diff --git a/src/asahi/vulkan/hk_cmd_buffer.c b/src/asahi/vulkan/hk_cmd_buffer.c
index 02eb976f336..cc88e79a0de 100644
--- a/src/asahi/vulkan/hk_cmd_buffer.c
+++ b/src/asahi/vulkan/hk_cmd_buffer.c
@@ -682,7 +682,7 @@ hk_upload_usc_words(struct hk_cmd_buffer *cmd, struct hk_shader *s,
 
    uint64_t root_ptr;
 
-   if (sw_stage == PIPE_SHADER_COMPUTE) {
+   if (sw_stage == MESA_SHADER_COMPUTE) {
       root_ptr = hk_cmd_buffer_upload_root(cmd, VK_PIPELINE_BIND_POINT_COMPUTE);
    } else {
       root_ptr = cmd->state.gfx.root;
diff --git a/src/compiler/shader_enums.h b/src/compiler/shader_enums.h
index 9a8ed438e77..6bb8feeb6b9 100644
--- a/src/compiler/shader_enums.h
+++ b/src/compiler/shader_enums.h
@@ -65,7 +65,7 @@ typedef enum pipe_shader_type
    MESA_SHADER_COMPUTE = 5,
    PIPE_SHADER_COMPUTE = MESA_SHADER_COMPUTE,
 
-   PIPE_SHADER_TYPES = (PIPE_SHADER_COMPUTE + 1),
+   PIPE_SHADER_TYPES = (MESA_SHADER_COMPUTE + 1),
    /* Vulkan-only stages. */
    MESA_SHADER_TASK         = 6,
    PIPE_SHADER_TASK = MESA_SHADER_TASK,
diff --git a/src/gallium/auxiliary/cso_cache/cso_context.c b/src/gallium/auxiliary/cso_cache/cso_context.c
index 443904e12a7..0dc8714b657 100644
--- a/src/gallium/auxiliary/cso_cache/cso_context.c
+++ b/src/gallium/auxiliary/cso_cache/cso_context.c
@@ -318,9 +318,9 @@ cso_create_context(struct pipe_context *pipe, unsigned flags)
    if (pipe->screen->shader_caps[MESA_SHADER_TESS_CTRL].max_instructions > 0) {
       ctx->has_tessellation = true;
    }
-   if (pipe->screen->shader_caps[PIPE_SHADER_COMPUTE].max_instructions > 0) {
+   if (pipe->screen->shader_caps[MESA_SHADER_COMPUTE].max_instructions > 0) {
       int supported_irs =
-         pipe->screen->shader_caps[PIPE_SHADER_COMPUTE].supported_irs;
+         pipe->screen->shader_caps[MESA_SHADER_COMPUTE].supported_irs;
       if (supported_irs & ((1 << PIPE_SHADER_IR_TGSI) |
                            (1 << PIPE_SHADER_IR_NIR))) {
          ctx->has_compute_shader = true;
@@ -375,7 +375,7 @@ cso_unbind_context(struct cso_context *cso)
                if (!ctx->has_tessellation)
                   continue;
                break;
-            case PIPE_SHADER_COMPUTE:
+            case MESA_SHADER_COMPUTE:
                if (!ctx->has_compute_shader)
                   continue;
                break;
@@ -1141,7 +1141,7 @@ cso_restore_compute_shader(struct cso_context_priv *ctx)
 static void
 cso_save_compute_samplers(struct cso_context_priv *ctx)
 {
-   struct sampler_info *info = &ctx->samplers[PIPE_SHADER_COMPUTE];
+   struct sampler_info *info = &ctx->samplers[MESA_SHADER_COMPUTE];
    struct sampler_info *saved = &ctx->compute_samplers_saved;
 
    memcpy(saved->cso_samplers, info->cso_samplers,
@@ -1153,7 +1153,7 @@ cso_save_compute_samplers(struct cso_context_priv *ctx)
 static void
 cso_restore_compute_samplers(struct cso_context_priv *ctx)
 {
-   struct sampler_info *info = &ctx->samplers[PIPE_SHADER_COMPUTE];
+   struct sampler_info *info = &ctx->samplers[MESA_SHADER_COMPUTE];
    struct sampler_info *saved = &ctx->compute_samplers_saved;
 
    memcpy(info->cso_samplers, saved->cso_samplers,
@@ -1167,7 +1167,7 @@ cso_restore_compute_samplers(struct cso_context_priv *ctx)
       }
    }
 
-   cso_single_sampler_done(&ctx->base, PIPE_SHADER_COMPUTE);
+   cso_single_sampler_done(&ctx->base, MESA_SHADER_COMPUTE);
 }
 
 
diff --git a/src/gallium/auxiliary/driver_ddebug/dd_draw.c b/src/gallium/auxiliary/driver_ddebug/dd_draw.c
index cc7d3ec4902..632be94d072 100644
--- a/src/gallium/auxiliary/driver_ddebug/dd_draw.c
+++ b/src/gallium/auxiliary/driver_ddebug/dd_draw.c
@@ -269,7 +269,7 @@ dd_dump_shader(struct dd_draw_state *dstate, enum pipe_shader_type sh, FILE *f)
    shader_str[MESA_SHADER_TESS_EVAL] = "TESS_EVAL";
    shader_str[MESA_SHADER_GEOMETRY] = "GEOMETRY";
    shader_str[MESA_SHADER_FRAGMENT] = "FRAGMENT";
-   shader_str[PIPE_SHADER_COMPUTE] = "COMPUTE";
+   shader_str[MESA_SHADER_COMPUTE] = "COMPUTE";
 
    if (sh == MESA_SHADER_TESS_CTRL &&
        !dstate->shaders[MESA_SHADER_TESS_CTRL] &&
@@ -404,7 +404,7 @@ dd_dump_draw_vbo(struct dd_draw_state *dstate, struct pipe_draw_info *info,
 
    fprintf(f, "\n");
    for (sh = 0; sh < PIPE_SHADER_TYPES; sh++) {
-      if (sh == PIPE_SHADER_COMPUTE)
+      if (sh == MESA_SHADER_COMPUTE)
          continue;
 
       dd_dump_shader(dstate, sh, f);
@@ -446,7 +446,7 @@ dd_dump_launch_grid(struct dd_draw_state *dstate, struct pipe_grid_info *info, F
    DUMP(grid_info, info);
    fprintf(f, "\n");
 
-   dd_dump_shader(dstate, PIPE_SHADER_COMPUTE, f);
+   dd_dump_shader(dstate, MESA_SHADER_COMPUTE, f);
    fprintf(f, "\n");
 }
 
diff --git a/src/gallium/auxiliary/tgsi/tgsi_exec.c b/src/gallium/auxiliary/tgsi/tgsi_exec.c
index e35f817f6f2..9acf51a1fb8 100644
--- a/src/gallium/auxiliary/tgsi/tgsi_exec.c
+++ b/src/gallium/auxiliary/tgsi/tgsi_exec.c
@@ -1220,7 +1220,7 @@ tgsi_exec_machine_create(enum pipe_shader_type shader_type)
 
    mach->ShaderType = shader_type;
 
-   if (shader_type != PIPE_SHADER_COMPUTE) {
+   if (shader_type != MESA_SHADER_COMPUTE) {
       mach->Inputs = align_malloc(sizeof(struct tgsi_exec_vector) * PIPE_MAX_SHADER_INPUTS, 16);
       mach->Outputs = align_malloc(sizeof(struct tgsi_exec_vector) * PIPE_MAX_SHADER_OUTPUTS, 16);
       if (!mach->Inputs || !mach->Outputs)
@@ -6028,7 +6028,7 @@ tgsi_exec_machine_run( struct tgsi_exec_machine *mach, int start_pc )
          barrier_hit = exec_instruction(mach, mach->Instructions + mach->pc, &mach->pc);
 
          /* for compute shaders if we hit a barrier return now for later rescheduling */
-         if (barrier_hit && mach->ShaderType == PIPE_SHADER_COMPUTE)
+         if (barrier_hit && mach->ShaderType == MESA_SHADER_COMPUTE)
             return 0;
 
 #if DEBUG_EXECUTION
diff --git a/src/gallium/auxiliary/tgsi/tgsi_from_mesa.h b/src/gallium/auxiliary/tgsi/tgsi_from_mesa.h
index 668825364db..06a401fa731 100644
--- a/src/gallium/auxiliary/tgsi/tgsi_from_mesa.h
+++ b/src/gallium/auxiliary/tgsi/tgsi_from_mesa.h
@@ -66,7 +66,7 @@ pipe_shader_type_from_mesa(gl_shader_stage stage)
    STATIC_ASSERT((enum pipe_shader_type) MESA_SHADER_TESS_CTRL == MESA_SHADER_TESS_CTRL);
    STATIC_ASSERT((enum pipe_shader_type) MESA_SHADER_TESS_EVAL == MESA_SHADER_TESS_EVAL);
    STATIC_ASSERT((enum pipe_shader_type) MESA_SHADER_GEOMETRY == MESA_SHADER_GEOMETRY);
-   STATIC_ASSERT((enum pipe_shader_type) MESA_SHADER_COMPUTE == PIPE_SHADER_COMPUTE);
+   STATIC_ASSERT((enum pipe_shader_type) MESA_SHADER_COMPUTE == MESA_SHADER_COMPUTE);
    return (enum pipe_shader_type)stage;
 }
 
diff --git a/src/gallium/auxiliary/tgsi/tgsi_scan.c b/src/gallium/auxiliary/tgsi/tgsi_scan.c
index 08009c291c9..8a1fcbcf821 100644
--- a/src/gallium/auxiliary/tgsi/tgsi_scan.c
+++ b/src/gallium/auxiliary/tgsi/tgsi_scan.c
@@ -85,7 +85,7 @@ scan_src_operand(struct tgsi_shader_info *info,
 {
    int ind = src->Register.Index;
 
-   if (info->processor == PIPE_SHADER_COMPUTE &&
+   if (info->processor == MESA_SHADER_COMPUTE &&
        src->Register.File == TGSI_FILE_SYSTEM_VALUE) {
       unsigned name;
 
@@ -615,7 +615,7 @@ tgsi_scan_shader(const struct tgsi_token *tokens,
           procType == MESA_SHADER_GEOMETRY ||
           procType == MESA_SHADER_TESS_CTRL ||
           procType == MESA_SHADER_TESS_EVAL ||
-          procType == PIPE_SHADER_COMPUTE);
+          procType == MESA_SHADER_COMPUTE);
    info->processor = procType;
 
    if (procType == MESA_SHADER_GEOMETRY)
diff --git a/src/gallium/auxiliary/tgsi/tgsi_text.c b/src/gallium/auxiliary/tgsi/tgsi_text.c
index c9ce8060c8d..c6126430d85 100644
--- a/src/gallium/auxiliary/tgsi/tgsi_text.c
+++ b/src/gallium/auxiliary/tgsi/tgsi_text.c
@@ -343,7 +343,7 @@ static bool parse_header( struct translate_ctx *ctx )
    else if (str_match_nocase_whole( &ctx->cur, "TESS_EVAL" ))
       processor = MESA_SHADER_TESS_EVAL;
    else if (str_match_nocase_whole( &ctx->cur, "COMP" ))
-      processor = PIPE_SHADER_COMPUTE;
+      processor = MESA_SHADER_COMPUTE;
    else {
       report_error( ctx, "Unknown header" );
       return false;
diff --git a/src/gallium/auxiliary/util/u_tests.c b/src/gallium/auxiliary/util/u_tests.c
index f93a9010839..90b8f67d4d6 100644
--- a/src/gallium/auxiliary/util/u_tests.c
+++ b/src/gallium/auxiliary/util/u_tests.c
@@ -843,7 +843,7 @@ test_compute_clear_image_shader(struct pipe_context *ctx)
    image.shader_access = image.access = PIPE_IMAGE_ACCESS_READ_WRITE;
    image.format = cb->format;
 
-   ctx->set_shader_images(ctx, PIPE_SHADER_COMPUTE, 0, 1, 0, &image);
+   ctx->set_shader_images(ctx, MESA_SHADER_COMPUTE, 0, 1, 0, &image);
 
    /* Dispatch compute. */
    struct pipe_grid_info info = {0};
diff --git a/src/gallium/auxiliary/util/u_threaded_context.c b/src/gallium/auxiliary/util/u_threaded_context.c
index dc6e2e32d43..e8ed1950366 100644
--- a/src/gallium/auxiliary/util/u_threaded_context.c
+++ b/src/gallium/auxiliary/util/u_threaded_context.c
@@ -932,7 +932,7 @@ tc_add_all_compute_bindings_to_buffer_list(struct threaded_context *tc)
 {
    BITSET_WORD *buffer_list = tc->buffer_lists[tc->next_buf_list].buffer_list;
 
-   tc_add_shader_bindings_to_buffer_list(tc, buffer_list, PIPE_SHADER_COMPUTE);
+   tc_add_shader_bindings_to_buffer_list(tc, buffer_list, MESA_SHADER_COMPUTE);
    tc->add_all_compute_bindings_to_buffer_list = false;
 }
 
@@ -964,7 +964,7 @@ tc_rebind_buffer(struct threaded_context *tc, uint32_t old_id, uint32_t new_id,
    if (tc->seen_gs)
       rebound += tc_rebind_shader_bindings(tc, old_id, new_id, MESA_SHADER_GEOMETRY, rebind_mask);
 
-   rebound += tc_rebind_shader_bindings(tc, old_id, new_id, PIPE_SHADER_COMPUTE, rebind_mask);
+   rebound += tc_rebind_shader_bindings(tc, old_id, new_id, MESA_SHADER_COMPUTE, rebind_mask);
 
    if (rebound)
       BITSET_SET(tc->buffer_lists[tc->next_buf_list].buffer_list, new_id & TC_BUFFER_ID_MASK);
@@ -1008,7 +1008,7 @@ tc_is_buffer_bound_for_write(struct threaded_context *tc, uint32_t id)
 
    if (tc_is_buffer_shader_bound_for_write(tc, id, MESA_SHADER_VERTEX) ||
        tc_is_buffer_shader_bound_for_write(tc, id, MESA_SHADER_FRAGMENT) ||
-       tc_is_buffer_shader_bound_for_write(tc, id, PIPE_SHADER_COMPUTE))
+       tc_is_buffer_shader_bound_for_write(tc, id, MESA_SHADER_COMPUTE))
       return true;
 
    if (tc->seen_tcs &&
diff --git a/src/gallium/auxiliary/vl/vl_compositor_cs.c b/src/gallium/auxiliary/vl/vl_compositor_cs.c
index a98d484da39..f7ce3874008 100644
--- a/src/gallium/auxiliary/vl/vl_compositor_cs.c
+++ b/src/gallium/auxiliary/vl/vl_compositor_cs.c
@@ -100,7 +100,7 @@ static nir_def *cs_create_shader(struct vl_compositor *c, struct cs_shader *s)
       glsl_sampler_type(sampler_dim, /*is_shadow*/ false, s->array, GLSL_TYPE_FLOAT);
    const struct glsl_type *image_type =
       glsl_image_type(GLSL_SAMPLER_DIM_2D, /*is_array*/ false, GLSL_TYPE_FLOAT);
-   const nir_shader_compiler_options *options = c->pipe->screen->nir_options[PIPE_SHADER_COMPUTE];
+   const nir_shader_compiler_options *options = c->pipe->screen->nir_options[MESA_SHADER_COMPUTE];
 
    s->b = nir_builder_init_simple_shader(MESA_SHADER_COMPUTE, options, "vl:%s", s->name);
    nir_builder *b = &s->b;
@@ -615,7 +615,7 @@ cs_launch(struct vl_compositor *c,
    image.shader_access = image.access = PIPE_IMAGE_ACCESS_READ_WRITE;
    image.format = c->fb_state.cbufs[0].texture->format;
 
-   ctx->set_shader_images(c->pipe, PIPE_SHADER_COMPUTE, 0, 1, 0, &image);
+   ctx->set_shader_images(c->pipe, MESA_SHADER_COMPUTE, 0, 1, 0, &image);
 
    /* Bind compute shader */
    ctx->bind_compute_state(ctx, cs);
@@ -854,20 +854,20 @@ draw_layers(struct vl_compositor       *c,
          calc_proj(layer, sampler1->texture, drawn.chroma_proj);
          set_viewport(s, &drawn, samplers);
 
-         c->pipe->bind_sampler_states(c->pipe, PIPE_SHADER_COMPUTE, 0,
+         c->pipe->bind_sampler_states(c->pipe, MESA_SHADER_COMPUTE, 0,
                         num_sampler_views, layer->samplers);
-         c->pipe->set_sampler_views(c->pipe, PIPE_SHADER_COMPUTE, 0,
+         c->pipe->set_sampler_views(c->pipe, MESA_SHADER_COMPUTE, 0,
                         num_sampler_views, 0, samplers);
 
          cs_launch(c, layer->cs, &(drawn.area));
 
          /* Unbind. */
-         c->pipe->set_shader_images(c->pipe, PIPE_SHADER_COMPUTE, 0, 0, 1, NULL);
-         c->pipe->set_constant_buffer(c->pipe, PIPE_SHADER_COMPUTE, 0, false, NULL);
-         c->pipe->set_sampler_views(c->pipe, PIPE_SHADER_COMPUTE, 0, 0,
+         c->pipe->set_shader_images(c->pipe, MESA_SHADER_COMPUTE, 0, 0, 1, NULL);
+         c->pipe->set_constant_buffer(c->pipe, MESA_SHADER_COMPUTE, 0, false, NULL);
+         c->pipe->set_sampler_views(c->pipe, MESA_SHADER_COMPUTE, 0, 0,
                         num_sampler_views, NULL);
          c->pipe->bind_compute_state(c->pipe, NULL);
-         c->pipe->bind_sampler_states(c->pipe, PIPE_SHADER_COMPUTE, 0,
+         c->pipe->bind_sampler_states(c->pipe, MESA_SHADER_COMPUTE, 0,
                         num_sampler_views, NULL);
 
          if (dirty) {
@@ -910,7 +910,7 @@ vl_compositor_cs_render(struct vl_compositor_state *s,
       dirty_area->x1 = dirty_area->y1 = VL_COMPOSITOR_MIN_DIRTY;
    }
 
-   pipe_set_constant_buffer(c->pipe, PIPE_SHADER_COMPUTE, 0, s->shader_params);
+   pipe_set_constant_buffer(c->pipe, MESA_SHADER_COMPUTE, 0, s->shader_params);
 
    draw_layers(c, s, dirty_area);
 }
diff --git a/src/gallium/auxiliary/vl/vl_deint_filter_cs.c b/src/gallium/auxiliary/vl/vl_deint_filter_cs.c
index 7844b69bf0c..4291ccba614 100644
--- a/src/gallium/auxiliary/vl/vl_deint_filter_cs.c
+++ b/src/gallium/auxiliary/vl/vl_deint_filter_cs.c
@@ -33,7 +33,7 @@ create_deint_shader(struct vl_deint_filter *filter, unsigned field)
       glsl_sampler_type(GLSL_SAMPLER_DIM_RECT, false, false, GLSL_TYPE_FLOAT);
    const struct glsl_type *image_type =
       glsl_image_type(GLSL_SAMPLER_DIM_2D, false, GLSL_TYPE_FLOAT);
-   const nir_shader_compiler_options *options = filter->pipe->screen->nir_options[PIPE_SHADER_COMPUTE];
+   const nir_shader_compiler_options *options = filter->pipe->screen->nir_options[MESA_SHADER_COMPUTE];
 
    nir_builder b = nir_builder_init_simple_shader(MESA_SHADER_COMPUTE, options, "vl:deint");
    b.shader->info.workgroup_size[0] = 8;
@@ -213,7 +213,7 @@ vl_deint_filter_cs_render(struct vl_deint_filter *filter,
    prev_sv = prev->get_sampler_view_planes(prev);
    next_sv = next->get_sampler_view_planes(next);
 
-   filter->pipe->bind_sampler_states(filter->pipe, PIPE_SHADER_COMPUTE,
+   filter->pipe->bind_sampler_states(filter->pipe, MESA_SHADER_COMPUTE,
                                      0, 4, filter->sampler);
 
    for (unsigned i = 0; i < 2; i++) {
@@ -224,7 +224,7 @@ vl_deint_filter_cs_render(struct vl_deint_filter *filter,
       sampler_views[1] = prev_sv[i];
       sampler_views[2] = cur_sv[i];
       sampler_views[3] = next_sv[i];
-      filter->pipe->set_sampler_views(filter->pipe, PIPE_SHADER_COMPUTE,
+      filter->pipe->set_sampler_views(filter->pipe, MESA_SHADER_COMPUTE,
                                       0, 4, 0, sampler_views);
 
       /* Bind the image */
@@ -234,7 +234,7 @@ vl_deint_filter_cs_render(struct vl_deint_filter *filter,
          .shader_access = PIPE_IMAGE_ACCESS_WRITE,
          .format = dst->texture->format,
       };
-      filter->pipe->set_shader_images(filter->pipe, PIPE_SHADER_COMPUTE,
+      filter->pipe->set_shader_images(filter->pipe, MESA_SHADER_COMPUTE,
                                       0, 1, 0, &image);
 
       /* Bind compute shader */
diff --git a/src/gallium/drivers/asahi/agx_blit.c b/src/gallium/drivers/asahi/agx_blit.c
index 230adf023b8..7e11fdae7f7 100644
--- a/src/gallium/drivers/asahi/agx_blit.c
+++ b/src/gallium/drivers/asahi/agx_blit.c
@@ -57,7 +57,7 @@ static void *
 asahi_blit_compute_shader(struct pipe_context *ctx, struct asahi_blit_key *key)
 {
    const nir_shader_compiler_options *options =
-      ctx->screen->nir_options[PIPE_SHADER_COMPUTE];
+      ctx->screen->nir_options[MESA_SHADER_COMPUTE];
 
    nir_builder b_ =
       nir_builder_init_simple_shader(MESA_SHADER_COMPUTE, options, "blit_cs");
@@ -244,7 +244,7 @@ static void
 asahi_compute_save(struct agx_context *ctx)
 {
    struct asahi_blitter *blitter = &ctx->compute_blitter;
-   struct agx_stage *stage = &ctx->stage[PIPE_SHADER_COMPUTE];
+   struct agx_stage *stage = &ctx->stage[MESA_SHADER_COMPUTE];
 
    assert(!blitter->active && "recursion detected, driver bug");
 
@@ -278,25 +278,25 @@ asahi_compute_restore(struct agx_context *ctx)
    struct asahi_blitter *blitter = &ctx->compute_blitter;
 
    if (blitter->has_saved_image) {
-      pctx->set_shader_images(pctx, PIPE_SHADER_COMPUTE, 0, 1, 0,
+      pctx->set_shader_images(pctx, MESA_SHADER_COMPUTE, 0, 1, 0,
                               &blitter->saved_image);
       pipe_resource_reference(&blitter->saved_image.resource, NULL);
    }
 
    /* take_ownership=true so do not unreference */
-   pctx->set_constant_buffer(pctx, PIPE_SHADER_COMPUTE, 0, true,
+   pctx->set_constant_buffer(pctx, MESA_SHADER_COMPUTE, 0, true,
                              &blitter->saved_cb);
    blitter->saved_cb.buffer = NULL;
 
    if (blitter->saved_sampler_view) {
-      pctx->set_sampler_views(pctx, PIPE_SHADER_COMPUTE, 0, 1, 0,
+      pctx->set_sampler_views(pctx, MESA_SHADER_COMPUTE, 0, 1, 0,
                               &blitter->saved_sampler_view);
 
       blitter->saved_sampler_view = NULL;
    }
 
    if (blitter->saved_num_sampler_states) {
-      pctx->bind_sampler_states(pctx, PIPE_SHADER_COMPUTE, 0,
+      pctx->bind_sampler_states(pctx, MESA_SHADER_COMPUTE, 0,
                                 blitter->saved_num_sampler_states,
                                 blitter->saved_sampler_states);
    }
@@ -366,7 +366,7 @@ asahi_compute_blit(struct pipe_context *ctx, const struct pipe_blit_info *info,
       .buffer_size = sizeof(data),
       .user_buffer = data,
    };
-   ctx->set_constant_buffer(ctx, PIPE_SHADER_COMPUTE, 0, false, &cb);
+   ctx->set_constant_buffer(ctx, MESA_SHADER_COMPUTE, 0, false, &cb);
 
    struct pipe_image_view image = {
       .resource = dst,
@@ -378,7 +378,7 @@ asahi_compute_blit(struct pipe_context *ctx, const struct pipe_blit_info *info,
       .u.tex.last_layer = info->dst.box.z + depth - 1,
       .u.tex.single_layer_view = !array,
    };
-   ctx->set_shader_images(ctx, PIPE_SHADER_COMPUTE, 0, 1, 0, &image);
+   ctx->set_shader_images(ctx, MESA_SHADER_COMPUTE, 0, 1, 0, &image);
 
    if (!blitter->sampler[info->filter]) {
       struct pipe_sampler_state sampler_state = {
@@ -396,7 +396,7 @@ asahi_compute_blit(struct pipe_context *ctx, const struct pipe_blit_info *info,
          ctx->create_sampler_state(ctx, &sampler_state);
    }
 
-   ctx->bind_sampler_states(ctx, PIPE_SHADER_COMPUTE, 0, 1,
+   ctx->bind_sampler_states(ctx, MESA_SHADER_COMPUTE, 0, 1,
                             &blitter->sampler[info->filter]);
 
    /* Initialize the sampler view. */
@@ -412,7 +412,7 @@ asahi_compute_blit(struct pipe_context *ctx, const struct pipe_blit_info *info,
    src_templ.u.tex.first_level = info->src.level;
    src_templ.u.tex.last_level = info->src.level;
    src_view = ctx->create_sampler_view(ctx, src, &src_templ);
-   ctx->set_sampler_views(ctx, PIPE_SHADER_COMPUTE, 0, 1, 0, &src_view);
+   ctx->set_sampler_views(ctx, MESA_SHADER_COMPUTE, 0, 1, 0, &src_view);
    ctx->sampler_view_release(ctx, src_view);
 
    struct asahi_blit_key key = {
@@ -447,9 +447,9 @@ asahi_compute_blit(struct pipe_context *ctx, const struct pipe_blit_info *info,
          },
    };
    ctx->launch_grid(ctx, &grid_info);
-   ctx->set_shader_images(ctx, PIPE_SHADER_COMPUTE, 0, 0, 1, NULL);
-   ctx->set_constant_buffer(ctx, PIPE_SHADER_COMPUTE, 0, false, NULL);
-   ctx->set_sampler_views(ctx, PIPE_SHADER_COMPUTE, 0, 0, 1, NULL);
+   ctx->set_shader_images(ctx, MESA_SHADER_COMPUTE, 0, 0, 1, NULL);
+   ctx->set_constant_buffer(ctx, MESA_SHADER_COMPUTE, 0, false, NULL);
+   ctx->set_sampler_views(ctx, MESA_SHADER_COMPUTE, 0, 0, 1, NULL);
 
    asahi_compute_restore(agx_context(ctx));
 }
diff --git a/src/gallium/drivers/asahi/agx_nir_lower_sysvals.c b/src/gallium/drivers/asahi/agx_nir_lower_sysvals.c
index fcbc1da01e9..b2571134204 100644
--- a/src/gallium/drivers/asahi/agx_nir_lower_sysvals.c
+++ b/src/gallium/drivers/asahi/agx_nir_lower_sysvals.c
@@ -411,7 +411,7 @@ lay_out_uniforms(struct agx_compiled_shader *shader, struct state *state)
          .length = 4,
       };
 
-      bool sw = state->hw_stage == PIPE_SHADER_COMPUTE;
+      bool sw = state->hw_stage == MESA_SHADER_COMPUTE;
       if (sw) {
          shader->push[shader->push_range_count++] = (struct agx_push_range){
             .uniform = AGX_ABI_VUNI_INPUT_ASSEMBLY(count),
diff --git a/src/gallium/drivers/asahi/agx_pipe.c b/src/gallium/drivers/asahi/agx_pipe.c
index bf215cbc32a..6bec0508da3 100644
--- a/src/gallium/drivers/asahi/agx_pipe.c
+++ b/src/gallium/drivers/asahi/agx_pipe.c
@@ -1864,7 +1864,7 @@ agx_init_shader_caps(struct pipe_screen *pscreen)
 {
    bool is_no16 = agx_device(pscreen)->debug & AGX_DBG_NO16;
 
-   for (unsigned i = 0; i <= PIPE_SHADER_COMPUTE; i++) {
+   for (unsigned i = 0; i <= MESA_SHADER_COMPUTE; i++) {
       struct pipe_shader_caps *caps =
          (struct pipe_shader_caps *)&pscreen->shader_caps[i];
 
diff --git a/src/gallium/drivers/asahi/agx_state.c b/src/gallium/drivers/asahi/agx_state.c
index d23050ef70c..63e4fc54f91 100644
--- a/src/gallium/drivers/asahi/agx_state.c
+++ b/src/gallium/drivers/asahi/agx_state.c
@@ -1658,7 +1658,7 @@ agx_compile_variant(struct agx_device *dev, struct pipe_context *pctx,
 
    if (pre_gs) {
       compiled->pre_gs =
-         agx_compile_nir(dev, pre_gs, &pctx->debug, PIPE_SHADER_COMPUTE, false,
+         agx_compile_nir(dev, pre_gs, &pctx->debug, MESA_SHADER_COMPUTE, false,
                          true, false, 0, NULL);
    }
 
@@ -2439,7 +2439,7 @@ agx_bind_tes_state(struct pipe_context *pctx, void *cso)
 static void
 agx_bind_cs_state(struct pipe_context *pctx, void *cso)
 {
-   agx_bind_shader_state(pctx, cso, PIPE_SHADER_COMPUTE);
+   agx_bind_shader_state(pctx, cso, MESA_SHADER_COMPUTE);
 }
 
 /* Forward declare because of the recursion hit with geometry shaders */
@@ -2605,7 +2605,7 @@ agx_build_meta_shader_internal(struct agx_context *ctx,
    }
 
    struct agx_compiled_shader *shader = agx_compile_nir(
-      dev, b.shader, NULL, PIPE_SHADER_COMPUTE, internal_kernel,
+      dev, b.shader, NULL, MESA_SHADER_COMPUTE, internal_kernel,
       !prolog && !(b.shader->info.stage == MESA_SHADER_FRAGMENT &&
                    b.shader->info.fs.uses_sample_shading),
       prolog || epilog, cf_base, NULL);
@@ -3056,7 +3056,7 @@ agx_launch_precomp(struct agx_batch *batch, struct agx_grid grid,
 
    agx_batch_add_bo(batch, cs->bo);
    agx_launch_internal(batch, grid, cs->b.workgroup, cs->b.launch,
-                       PIPE_SHADER_COMPUTE, usc);
+                       MESA_SHADER_COMPUTE, usc);
 }
 
 struct asahi_bg_eot
@@ -4170,7 +4170,7 @@ agx_launch_gs_prerast(struct agx_batch *batch,
    if (xfb_or_queries) {
       perf_debug(dev, "Geometry shader transform feedback / query program");
       agx_launch(batch, agx_1d(1), agx_workgroup(1, 1, 1), gs->pre_gs, NULL,
-                 PIPE_SHADER_COMPUTE, 0);
+                 MESA_SHADER_COMPUTE, 0);
    }
 
    /* Pre-rast geometry shader */
@@ -5353,7 +5353,7 @@ agx_launch(struct agx_batch *batch, struct agx_grid grid,
    }
 #endif
 
-   uint32_t usc = agx_build_pipeline(batch, cs, linked, PIPE_SHADER_COMPUTE,
+   uint32_t usc = agx_build_pipeline(batch, cs, linked, MESA_SHADER_COMPUTE,
                                      variable_shared_mem, subgroups_per_core);
 
    if (cs)
@@ -5418,7 +5418,7 @@ agx_launch_grid(struct pipe_context *pipe, const struct pipe_grid_info *info)
    agx_batch_init_state(batch);
 
    struct agx_uncompiled_shader *uncompiled =
-      ctx->stage[PIPE_SHADER_COMPUTE].shader;
+      ctx->stage[MESA_SHADER_COMPUTE].shader;
 
    /* There is exactly one variant, get it */
    struct agx_compiled_shader *cs =
@@ -5436,7 +5436,7 @@ agx_launch_grid(struct pipe_context *pipe, const struct pipe_grid_info *info)
       }
    }
 
-   agx_launch(batch, grid, wg, cs, NULL, PIPE_SHADER_COMPUTE,
+   agx_launch(batch, grid, wg, cs, NULL, MESA_SHADER_COMPUTE,
               info->variable_shared_mem);
 
    /* TODO: Dirty tracking? */
diff --git a/src/gallium/drivers/asahi/agx_state.h b/src/gallium/drivers/asahi/agx_state.h
index fb89517616e..f7daec71836 100644
--- a/src/gallium/drivers/asahi/agx_state.h
+++ b/src/gallium/drivers/asahi/agx_state.h
@@ -98,7 +98,7 @@ static_assert(AGX_SYSVAL_STAGE(MESA_SHADER_GEOMETRY) == AGX_SYSVAL_TABLE_GS,
               "fixed enum orderings");
 static_assert(AGX_SYSVAL_STAGE(MESA_SHADER_FRAGMENT) == AGX_SYSVAL_TABLE_FS,
               "fixed enum orderings");
-static_assert(AGX_SYSVAL_STAGE(PIPE_SHADER_COMPUTE) == AGX_SYSVAL_TABLE_CS,
+static_assert(AGX_SYSVAL_STAGE(MESA_SHADER_COMPUTE) == AGX_SYSVAL_TABLE_CS,
               "fixed enum orderings");
 
 /* Root system value table */
@@ -823,7 +823,7 @@ agx_dirty_reset_graphics(struct agx_context *ctx)
    ctx->dirty = 0;
 
    for (unsigned i = 0; i < ARRAY_SIZE(ctx->stage); ++i) {
-      if (i != PIPE_SHADER_COMPUTE)
+      if (i != MESA_SHADER_COMPUTE)
          ctx->stage[i].dirty = 0;
    }
 }
diff --git a/src/gallium/drivers/crocus/crocus_screen.c b/src/gallium/drivers/crocus/crocus_screen.c
index 1a8ac779e6a..f88834a983e 100644
--- a/src/gallium/drivers/crocus/crocus_screen.c
+++ b/src/gallium/drivers/crocus/crocus_screen.c
@@ -139,7 +139,7 @@ crocus_init_shader_caps(struct crocus_screen *screen)
 {
    const struct intel_device_info *devinfo = &screen->devinfo;
 
-   for (unsigned i = 0; i <= PIPE_SHADER_COMPUTE; i++) {
+   for (unsigned i = 0; i <= MESA_SHADER_COMPUTE; i++) {
       struct pipe_shader_caps *caps =
          (struct pipe_shader_caps *)&screen->base.shader_caps[i];
 
@@ -181,7 +181,7 @@ crocus_init_shader_caps(struct crocus_screen *screen)
          (devinfo->verx10 >= 75) ? CROCUS_MAX_TEXTURE_SAMPLERS : 16;
 
       if (devinfo->ver >= 7 &&
-          (i == MESA_SHADER_FRAGMENT || i == PIPE_SHADER_COMPUTE))
+          (i == MESA_SHADER_FRAGMENT || i == MESA_SHADER_COMPUTE))
          caps->max_shader_images = CROCUS_MAX_TEXTURE_SAMPLERS;
 
       caps->max_shader_buffers =
diff --git a/src/gallium/drivers/d3d12/d3d12_compiler.cpp b/src/gallium/drivers/d3d12/d3d12_compiler.cpp
index c01d1393e14..7cfdaca5b2f 100644
--- a/src/gallium/drivers/d3d12/d3d12_compiler.cpp
+++ b/src/gallium/drivers/d3d12/d3d12_compiler.cpp
@@ -734,7 +734,7 @@ d3d12_compare_shader_keys(struct d3d12_selection_context* sel_ctx, const d3d12_s
       if (expect->fs.all != have->fs.all)
          return false;
       break;
-   case PIPE_SHADER_COMPUTE:
+   case MESA_SHADER_COMPUTE:
       if (memcmp(expect->cs.workgroup_size, have->cs.workgroup_size,
                  sizeof(have->cs.workgroup_size)))
          return false;
@@ -811,7 +811,7 @@ d3d12_shader_key_hash(const d3d12_shader_key *key)
    case MESA_SHADER_FRAGMENT:
       hash += key->fs.all;
       break;
-   case PIPE_SHADER_COMPUTE:
+   case MESA_SHADER_COMPUTE:
       hash = _mesa_hash_data_with_seed(&key->cs, sizeof(key->cs), hash);
       break;
    case MESA_SHADER_TESS_CTRL:
@@ -859,7 +859,7 @@ d3d12_fill_shader_key(struct d3d12_selection_context *sel_ctx,
       key->ds.tcs_vertices_out = 0;
       key->ds.prev_patch_outputs = 0;
       break;
-   case PIPE_SHADER_COMPUTE:
+   case MESA_SHADER_COMPUTE:
       memset(key->cs.workgroup_size, 0, sizeof(key->cs.workgroup_size));
       break;
    default: UNREACHABLE("Invalid stage type");
@@ -1036,7 +1036,7 @@ d3d12_fill_shader_key(struct d3d12_selection_context *sel_ctx,
       key->fs.remap_front_facing = 1;
    }
 
-   if (stage == PIPE_SHADER_COMPUTE && sel_ctx->variable_workgroup_size) {
+   if (stage == MESA_SHADER_COMPUTE && sel_ctx->variable_workgroup_size) {
       memcpy(key->cs.workgroup_size, sel_ctx->variable_workgroup_size, sizeof(key->cs.workgroup_size));
    }
 
@@ -1137,7 +1137,7 @@ select_shader_variant(struct d3d12_selection_context *sel_ctx, d3d12_shader_sele
       NIR_PASS(_, new_nir_variant, d3d12_lower_image_casts, &image_format_arr);
    }
 
-   if (key.stage == PIPE_SHADER_COMPUTE && sel->workgroup_size_variable) {
+   if (key.stage == MESA_SHADER_COMPUTE && sel->workgroup_size_variable) {
       new_nir_variant->info.workgroup_size[0] = static_cast<uint16_t>(key.cs.workgroup_size[0]);
       new_nir_variant->info.workgroup_size[1] = static_cast<uint16_t>(key.cs.workgroup_size[1]);
       new_nir_variant->info.workgroup_size[2] = static_cast<uint16_t>(key.cs.workgroup_size[2]);
@@ -1464,7 +1464,7 @@ d3d12_create_compute_shader(struct d3d12_context *ctx,
                             const struct pipe_compute_state *shader)
 {
    struct d3d12_shader_selector *sel = rzalloc(nullptr, d3d12_shader_selector);
-   sel->stage = PIPE_SHADER_COMPUTE;
+   sel->stage = MESA_SHADER_COMPUTE;
 
    struct nir_shader *nir = NULL;
 
diff --git a/src/gallium/drivers/d3d12/d3d12_compute_transforms.cpp b/src/gallium/drivers/d3d12/d3d12_compute_transforms.cpp
index 8585eb673b1..7ce86dba08f 100644
--- a/src/gallium/drivers/d3d12/d3d12_compute_transforms.cpp
+++ b/src/gallium/drivers/d3d12/d3d12_compute_transforms.cpp
@@ -486,12 +486,12 @@ d3d12_save_compute_transform_state(struct d3d12_context *ctx, d3d12_compute_tran
    memset(save, 0, sizeof(*save));
    save->cs = ctx->compute_state;
 
-   pipe_resource_reference(&save->cbuf0.buffer, ctx->cbufs[PIPE_SHADER_COMPUTE][1].buffer);
-   save->cbuf0 = ctx->cbufs[PIPE_SHADER_COMPUTE][1];
+   pipe_resource_reference(&save->cbuf0.buffer, ctx->cbufs[MESA_SHADER_COMPUTE][1].buffer);
+   save->cbuf0 = ctx->cbufs[MESA_SHADER_COMPUTE][1];
 
    for (unsigned i = 0; i < ARRAY_SIZE(save->ssbos); ++i) {
-      pipe_resource_reference(&save->ssbos[i].buffer, ctx->ssbo_views[PIPE_SHADER_COMPUTE][i].buffer);
-      save->ssbos[i] = ctx->ssbo_views[PIPE_SHADER_COMPUTE][i];
+      pipe_resource_reference(&save->ssbos[i].buffer, ctx->ssbo_views[MESA_SHADER_COMPUTE][i].buffer);
+      save->ssbos[i] = ctx->ssbo_views[MESA_SHADER_COMPUTE][i];
    }
 
    save->queries_disabled = ctx->queries_disabled;
@@ -505,8 +505,8 @@ d3d12_restore_compute_transform_state(struct d3d12_context *ctx, d3d12_compute_t
 
    ctx->base.bind_compute_state(&ctx->base, save->cs);
 
-   ctx->base.set_constant_buffer(&ctx->base, PIPE_SHADER_COMPUTE, 1, true, &save->cbuf0);
-   ctx->base.set_shader_buffers(&ctx->base, PIPE_SHADER_COMPUTE, 0, ARRAY_SIZE(save->ssbos), save->ssbos, (1u << ARRAY_SIZE(save->ssbos)) - 1);
+   ctx->base.set_constant_buffer(&ctx->base, MESA_SHADER_COMPUTE, 1, true, &save->cbuf0);
+   ctx->base.set_shader_buffers(&ctx->base, MESA_SHADER_COMPUTE, 0, ARRAY_SIZE(save->ssbos), save->ssbos, (1u << ARRAY_SIZE(save->ssbos)) - 1);
 
    if (ctx->current_predication)
       d3d12_enable_predication(ctx);
diff --git a/src/gallium/drivers/d3d12/d3d12_context_graphics.cpp b/src/gallium/drivers/d3d12/d3d12_context_graphics.cpp
index f5a6e43ab62..3f979a836e4 100644
--- a/src/gallium/drivers/d3d12/d3d12_context_graphics.cpp
+++ b/src/gallium/drivers/d3d12/d3d12_context_graphics.cpp
@@ -1839,7 +1839,7 @@ d3d12_disable_fake_so_buffers(struct d3d12_context *ctx)
       new_cs_ssbos[1].buffer = target->fill_buffer;
       new_cs_ssbos[1].buffer_offset = target->fill_buffer_offset;
       new_cs_ssbos[1].buffer_size = target->fill_buffer->width0 - target->fill_buffer_offset;
-      ctx->base.set_shader_buffers(&ctx->base, PIPE_SHADER_COMPUTE, 0, 2, new_cs_ssbos, 2);
+      ctx->base.set_shader_buffers(&ctx->base, MESA_SHADER_COMPUTE, 0, 2, new_cs_ssbos, 2);
 
       pipe_grid_info grid = {};
       grid.block[0] = grid.block[1] = grid.block[2] = 1;
@@ -1875,13 +1875,13 @@ d3d12_disable_fake_so_buffers(struct d3d12_context *ctx)
       new_cs_ssbos[1].buffer = fake_target->base.buffer;
       new_cs_ssbos[1].buffer_offset = fake_target->base.buffer_offset;
       new_cs_ssbos[1].buffer_size = fake_target->base.buffer_size;
-      ctx->base.set_shader_buffers(&ctx->base, PIPE_SHADER_COMPUTE, 0, 2, new_cs_ssbos, 2);
+      ctx->base.set_shader_buffers(&ctx->base, MESA_SHADER_COMPUTE, 0, 2, new_cs_ssbos, 2);
 
       pipe_constant_buffer cbuf = {};
       cbuf.buffer = fake_target->fill_buffer;
       cbuf.buffer_offset = fake_target->fill_buffer_offset;
       cbuf.buffer_size = fake_target->fill_buffer->width0 - cbuf.buffer_offset;
-      ctx->base.set_constant_buffer(&ctx->base, PIPE_SHADER_COMPUTE, 1, false, &cbuf);
+      ctx->base.set_constant_buffer(&ctx->base, MESA_SHADER_COMPUTE, 1, false, &cbuf);
 
       grid.indirect = fake_target->fill_buffer;
       grid.indirect_offset = fake_target->fill_buffer_offset + 4;
diff --git a/src/gallium/drivers/d3d12/d3d12_draw.cpp b/src/gallium/drivers/d3d12/d3d12_draw.cpp
index 97d5aa54fce..73de1871268 100644
--- a/src/gallium/drivers/d3d12/d3d12_draw.cpp
+++ b/src/gallium/drivers/d3d12/d3d12_draw.cpp
@@ -795,7 +795,7 @@ update_draw_indirect_with_sysvals(struct d3d12_context *ctx,
       draw_count_cbuf.buffer_offset = indirect_in->indirect_draw_count_offset;
       draw_count_cbuf.buffer_size = 4;
       draw_count_cbuf.user_buffer = nullptr;
-      ctx->base.set_constant_buffer(&ctx->base, PIPE_SHADER_COMPUTE, 1, false, &draw_count_cbuf);
+      ctx->base.set_constant_buffer(&ctx->base, MESA_SHADER_COMPUTE, 1, false, &draw_count_cbuf);
    }
    
    pipe_shader_buffer new_cs_ssbos[2];
@@ -815,7 +815,7 @@ update_draw_indirect_with_sysvals(struct d3d12_context *ctx,
    new_cs_ssbos[1].buffer = ctx->base.screen->resource_create(ctx->base.screen, &output_buf_templ);
    new_cs_ssbos[1].buffer_offset = 0;
    new_cs_ssbos[1].buffer_size = output_buf_templ.width0;
-   ctx->base.set_shader_buffers(&ctx->base, PIPE_SHADER_COMPUTE, 0, 2, new_cs_ssbos, 2);
+   ctx->base.set_shader_buffers(&ctx->base, MESA_SHADER_COMPUTE, 0, 2, new_cs_ssbos, 2);
 
    pipe_grid_info grid = {};
    grid.block[0] = grid.block[1] = grid.block[2] = 1;
@@ -863,7 +863,7 @@ update_draw_auto(struct d3d12_context *ctx,
    new_cs_ssbo.buffer = target->fill_buffer;
    new_cs_ssbo.buffer_offset = target->fill_buffer_offset;
    new_cs_ssbo.buffer_size = target->fill_buffer->width0 - new_cs_ssbo.buffer_offset;
-   ctx->base.set_shader_buffers(&ctx->base, PIPE_SHADER_COMPUTE, 0, 1, &new_cs_ssbo, 1);
+   ctx->base.set_shader_buffers(&ctx->base, MESA_SHADER_COMPUTE, 0, 1, &new_cs_ssbo, 1);
 
    pipe_grid_info grid = {};
    grid.block[0] = grid.block[1] = grid.block[2] = 1;
@@ -1352,7 +1352,7 @@ d3d12_launch_grid(struct pipe_context *pctx, const struct pipe_grid_info *info)
       if (ctx->compute_pipeline_state.root_signature != root_signature) {
          ctx->compute_pipeline_state.root_signature = root_signature;
          ctx->state_dirty |= D3D12_DIRTY_COMPUTE_ROOT_SIGNATURE;
-         ctx->shader_dirty[PIPE_SHADER_COMPUTE] |= D3D12_SHADER_DIRTY_ALL;
+         ctx->shader_dirty[MESA_SHADER_COMPUTE] |= D3D12_SHADER_DIRTY_ALL;
       }
    }
 
@@ -1414,6 +1414,6 @@ d3d12_launch_grid(struct pipe_context *pctx, const struct pipe_grid_info *info)
    ctx->cmdlist_dirty |= D3D12_DIRTY_SHADER;
    batch->pending_memory_barrier = false;
 
-   ctx->shader_dirty[PIPE_SHADER_COMPUTE] = 0;
+   ctx->shader_dirty[MESA_SHADER_COMPUTE] = 0;
    pipe_resource_reference(&patched_indirect, nullptr);
 }
diff --git a/src/gallium/drivers/d3d12/d3d12_query.cpp b/src/gallium/drivers/d3d12/d3d12_query.cpp
index 6708633ea65..e58a4f16b12 100644
--- a/src/gallium/drivers/d3d12/d3d12_query.cpp
+++ b/src/gallium/drivers/d3d12/d3d12_query.cpp
@@ -387,7 +387,7 @@ accumulate_subresult_gpu(struct d3d12_context *ctx, struct d3d12_query *q_parent
    new_cs_ssbos[0].buffer = q_parent->subqueries[sub_query].buffer;
    new_cs_ssbos[0].buffer_offset = q_parent->subqueries[sub_query].buffer_offset;
    new_cs_ssbos[0].buffer_size = q_parent->subqueries[sub_query].query_size * q_parent->subqueries[sub_query].num_queries;
-   ctx->base.set_shader_buffers(&ctx->base, PIPE_SHADER_COMPUTE, 0, 1, new_cs_ssbos, 1);
+   ctx->base.set_shader_buffers(&ctx->base, MESA_SHADER_COMPUTE, 0, 1, new_cs_ssbos, 1);
 
    pipe_grid_info grid = {};
    grid.block[0] = grid.block[1] = grid.block[2] = 1;
@@ -435,7 +435,7 @@ accumulate_result_gpu(struct d3d12_context *ctx, struct d3d12_query *q,
    new_cs_ssbos[num_ssbos].buffer_size = dst->width0;
    num_ssbos++;
    
-   ctx->base.set_shader_buffers(&ctx->base, PIPE_SHADER_COMPUTE, 0, num_ssbos, new_cs_ssbos, 1 << (num_ssbos - 1));
+   ctx->base.set_shader_buffers(&ctx->base, MESA_SHADER_COMPUTE, 0, num_ssbos, new_cs_ssbos, 1 << (num_ssbos - 1));
 
    pipe_grid_info grid = {};
    grid.block[0] = grid.block[1] = grid.block[2] = 1;
diff --git a/src/gallium/drivers/d3d12/d3d12_root_signature.cpp b/src/gallium/drivers/d3d12/d3d12_root_signature.cpp
index b3e85508ec2..7fd94f6f1dd 100644
--- a/src/gallium/drivers/d3d12/d3d12_root_signature.cpp
+++ b/src/gallium/drivers/d3d12/d3d12_root_signature.cpp
@@ -51,7 +51,7 @@ get_shader_visibility(enum pipe_shader_type stage)
       return D3D12_SHADER_VISIBILITY_HULL;
    case MESA_SHADER_TESS_EVAL:
       return D3D12_SHADER_VISIBILITY_DOMAIN;
-   case PIPE_SHADER_COMPUTE:
+   case MESA_SHADER_COMPUTE:
       return D3D12_SHADER_VISIBILITY_ALL;
    default:
       UNREACHABLE("unknown shader stage");
@@ -122,7 +122,7 @@ create_root_signature(struct d3d12_context *ctx, struct d3d12_root_signature_key
 
    unsigned count = key->compute ? 1 : D3D12_GFX_SHADER_STAGES;
    for (unsigned i = 0; i < count; ++i) {
-      unsigned stage = key->compute ? PIPE_SHADER_COMPUTE : i;
+      unsigned stage = key->compute ? MESA_SHADER_COMPUTE : i;
       D3D12_SHADER_VISIBILITY visibility = get_shader_visibility((enum pipe_shader_type)stage);
 
       if (key->stages[i].end_cb_bindings - key->stages[i].begin_cb_bindings > 0) {
diff --git a/src/gallium/drivers/d3d12/d3d12_screen.cpp b/src/gallium/drivers/d3d12/d3d12_screen.cpp
index 0b792674d0a..82669a18077 100644
--- a/src/gallium/drivers/d3d12/d3d12_screen.cpp
+++ b/src/gallium/drivers/d3d12/d3d12_screen.cpp
@@ -127,7 +127,7 @@ d3d12_get_video_mem(struct pipe_screen *pscreen)
 static void
 d3d12_init_shader_caps(struct d3d12_screen *screen)
 {
-   for (unsigned i = 0; i <= PIPE_SHADER_COMPUTE; i++) {
+   for (unsigned i = 0; i <= MESA_SHADER_COMPUTE; i++) {
       struct pipe_shader_caps *caps =
          (struct pipe_shader_caps *)&screen->base.shader_caps[i];
 
diff --git a/src/gallium/drivers/freedreno/a4xx/fd4_compute.c b/src/gallium/drivers/freedreno/a4xx/fd4_compute.c
index 42e94d1c8d8..a170c82c609 100644
--- a/src/gallium/drivers/freedreno/a4xx/fd4_compute.c
+++ b/src/gallium/drivers/freedreno/a4xx/fd4_compute.c
@@ -121,7 +121,7 @@ fd4_launch_grid(struct fd_context *ctx,
    if (!v)
       return;
 
-   if (ctx->dirty_shader[PIPE_SHADER_COMPUTE] & FD_DIRTY_SHADER_PROG)
+   if (ctx->dirty_shader[MESA_SHADER_COMPUTE] & FD_DIRTY_SHADER_PROG)
       cs_program_emit(ring, v);
 
    fd4_emit_cs_state(ctx, ring, v);
diff --git a/src/gallium/drivers/freedreno/a4xx/fd4_emit.c b/src/gallium/drivers/freedreno/a4xx/fd4_emit.c
index 51f99507ebf..9a8ac9a3f85 100644
--- a/src/gallium/drivers/freedreno/a4xx/fd4_emit.c
+++ b/src/gallium/drivers/freedreno/a4xx/fd4_emit.c
@@ -913,13 +913,13 @@ void
 fd4_emit_cs_state(struct fd_context *ctx, struct fd_ringbuffer *ring,
                   struct ir3_shader_variant *cp)
 {
-   enum fd_dirty_shader_state dirty = ctx->dirty_shader[PIPE_SHADER_COMPUTE];
-   unsigned num_textures = ctx->tex[PIPE_SHADER_COMPUTE].num_textures +
+   enum fd_dirty_shader_state dirty = ctx->dirty_shader[MESA_SHADER_COMPUTE];
+   unsigned num_textures = ctx->tex[MESA_SHADER_COMPUTE].num_textures +
       cp->astc_srgb.count +
       cp->tg4.count;
 
    if (dirty & FD_DIRTY_SHADER_TEX) {
-      emit_textures(ctx, ring, SB4_CS_TEX, &ctx->tex[PIPE_SHADER_COMPUTE], cp);
+      emit_textures(ctx, ring, SB4_CS_TEX, &ctx->tex[MESA_SHADER_COMPUTE], cp);
 
       OUT_PKT0(ring, REG_A4XX_TPL1_TP_TEX_COUNT, 1);
       OUT_RING(ring, 0);
@@ -927,13 +927,13 @@ fd4_emit_cs_state(struct fd_context *ctx, struct fd_ringbuffer *ring,
 
    OUT_PKT0(ring, REG_A4XX_TPL1_TP_FS_TEX_COUNT, 1);
    OUT_RING(ring, A4XX_TPL1_TP_FS_TEX_COUNT_CS(
-               ctx->shaderimg[PIPE_SHADER_COMPUTE].enabled_mask ? 0x80 : num_textures));
+               ctx->shaderimg[MESA_SHADER_COMPUTE].enabled_mask ? 0x80 : num_textures));
 
    if (dirty & FD_DIRTY_SHADER_SSBO)
-      emit_ssbos(ctx, ring, SB4_CS_SSBO, &ctx->shaderbuf[PIPE_SHADER_COMPUTE]);
+      emit_ssbos(ctx, ring, SB4_CS_SSBO, &ctx->shaderbuf[MESA_SHADER_COMPUTE]);
 
    if (dirty & FD_DIRTY_SHADER_IMAGE)
-      fd4_emit_images(ctx, ring, PIPE_SHADER_COMPUTE, cp);
+      fd4_emit_images(ctx, ring, MESA_SHADER_COMPUTE, cp);
 }
 
 /* emit setup at begin of new cmdstream buffer (don't rely on previous
diff --git a/src/gallium/drivers/freedreno/a4xx/fd4_image.c b/src/gallium/drivers/freedreno/a4xx/fd4_image.c
index 82cdd870030..74a28b488ce 100644
--- a/src/gallium/drivers/freedreno/a4xx/fd4_image.c
+++ b/src/gallium/drivers/freedreno/a4xx/fd4_image.c
@@ -14,12 +14,12 @@
 #include "fd4_texture.h"
 
 static enum a4xx_state_block texsb[] = {
-   [PIPE_SHADER_COMPUTE] = SB4_CS_TEX,
+   [MESA_SHADER_COMPUTE] = SB4_CS_TEX,
    [MESA_SHADER_FRAGMENT] = SB4_FS_TEX,
 };
 
 static enum a4xx_state_block imgsb[] = {
-   [PIPE_SHADER_COMPUTE] = SB4_CS_SSBO,
+   [MESA_SHADER_COMPUTE] = SB4_CS_SSBO,
    [MESA_SHADER_FRAGMENT] = SB4_SSBO,
 };
 
diff --git a/src/gallium/drivers/freedreno/a4xx/fd4_texture.c b/src/gallium/drivers/freedreno/a4xx/fd4_texture.c
index ef0df31888b..9a4986f6828 100644
--- a/src/gallium/drivers/freedreno/a4xx/fd4_texture.c
+++ b/src/gallium/drivers/freedreno/a4xx/fd4_texture.c
@@ -228,7 +228,7 @@ fd4_set_sampler_views(struct pipe_context *pctx, enum pipe_shader_type shader,
       sampler_swizzles = fd4_ctx->fsampler_swizzles;
    } else if (shader == MESA_SHADER_VERTEX) {
       sampler_swizzles = fd4_ctx->vsampler_swizzles;
-   } else if (shader == PIPE_SHADER_COMPUTE) {
+   } else if (shader == MESA_SHADER_COMPUTE) {
       sampler_swizzles = fd4_ctx->csampler_swizzles;
    } else {
       assert(0);
@@ -278,7 +278,7 @@ fd4_set_sampler_views(struct pipe_context *pctx, enum pipe_shader_type shader,
       fd4_ctx->fastc_srgb = astc_srgb;
    } else if (shader == MESA_SHADER_VERTEX) {
       fd4_ctx->vastc_srgb = astc_srgb;
-   } else if (shader == PIPE_SHADER_COMPUTE) {
+   } else if (shader == MESA_SHADER_COMPUTE) {
       fd4_ctx->castc_srgb = astc_srgb;
    }
 }
diff --git a/src/gallium/drivers/freedreno/a5xx/fd5_compute.c b/src/gallium/drivers/freedreno/a5xx/fd5_compute.c
index 443bc364853..4344074f58f 100644
--- a/src/gallium/drivers/freedreno/a5xx/fd5_compute.c
+++ b/src/gallium/drivers/freedreno/a5xx/fd5_compute.c
@@ -100,7 +100,7 @@ fd5_launch_grid(struct fd_context *ctx,
    if (!v)
       return;
 
-   if (ctx->dirty_shader[PIPE_SHADER_COMPUTE] & FD_DIRTY_SHADER_PROG)
+   if (ctx->dirty_shader[MESA_SHADER_COMPUTE] & FD_DIRTY_SHADER_PROG)
       cs_program_emit(ctx, ring, v);
 
    fd5_emit_cs_state(ctx, ring, v);
diff --git a/src/gallium/drivers/freedreno/a5xx/fd5_emit.c b/src/gallium/drivers/freedreno/a5xx/fd5_emit.c
index 24e0f400deb..862dbc6ec1b 100644
--- a/src/gallium/drivers/freedreno/a5xx/fd5_emit.c
+++ b/src/gallium/drivers/freedreno/a5xx/fd5_emit.c
@@ -859,12 +859,12 @@ void
 fd5_emit_cs_state(struct fd_context *ctx, struct fd_ringbuffer *ring,
                   struct ir3_shader_variant *cp)
 {
-   enum fd_dirty_shader_state dirty = ctx->dirty_shader[PIPE_SHADER_COMPUTE];
+   enum fd_dirty_shader_state dirty = ctx->dirty_shader[MESA_SHADER_COMPUTE];
 
    if (dirty & FD_DIRTY_SHADER_TEX) {
       bool needs_border = false;
       needs_border |=
-         emit_textures(ctx, ring, SB4_CS_TEX, &ctx->tex[PIPE_SHADER_COMPUTE]);
+         emit_textures(ctx, ring, SB4_CS_TEX, &ctx->tex[MESA_SHADER_COMPUTE]);
 
       if (needs_border)
          emit_border_color(ctx, ring);
@@ -886,16 +886,16 @@ fd5_emit_cs_state(struct fd_context *ctx, struct fd_ringbuffer *ring,
    }
 
    OUT_PKT4(ring, REG_A5XX_TPL1_CS_TEX_COUNT, 1);
-   OUT_RING(ring, ctx->shaderimg[PIPE_SHADER_COMPUTE].enabled_mask
+   OUT_RING(ring, ctx->shaderimg[MESA_SHADER_COMPUTE].enabled_mask
                      ? ~0
-                     : ctx->tex[PIPE_SHADER_COMPUTE].num_textures);
+                     : ctx->tex[MESA_SHADER_COMPUTE].num_textures);
 
    if (dirty & FD_DIRTY_SHADER_SSBO)
-      emit_ssbos(ctx, ring, SB4_CS_SSBO, &ctx->shaderbuf[PIPE_SHADER_COMPUTE],
+      emit_ssbos(ctx, ring, SB4_CS_SSBO, &ctx->shaderbuf[MESA_SHADER_COMPUTE],
                  cp);
 
    if (dirty & FD_DIRTY_SHADER_IMAGE)
-      fd5_emit_images(ctx, ring, PIPE_SHADER_COMPUTE, cp);
+      fd5_emit_images(ctx, ring, MESA_SHADER_COMPUTE, cp);
 }
 
 /* emit setup at begin of new cmdstream buffer (don't rely on previous
diff --git a/src/gallium/drivers/freedreno/a5xx/fd5_image.c b/src/gallium/drivers/freedreno/a5xx/fd5_image.c
index d732f7dd069..7fc0838673d 100644
--- a/src/gallium/drivers/freedreno/a5xx/fd5_image.c
+++ b/src/gallium/drivers/freedreno/a5xx/fd5_image.c
@@ -14,12 +14,12 @@
 #include "freedreno_resource.h"
 
 static enum a4xx_state_block texsb[] = {
-   [PIPE_SHADER_COMPUTE] = SB4_CS_TEX,
+   [MESA_SHADER_COMPUTE] = SB4_CS_TEX,
    [MESA_SHADER_FRAGMENT] = SB4_FS_TEX,
 };
 
 static enum a4xx_state_block imgsb[] = {
-   [PIPE_SHADER_COMPUTE] = SB4_CS_SSBO,
+   [MESA_SHADER_COMPUTE] = SB4_CS_SSBO,
    [MESA_SHADER_FRAGMENT] = SB4_SSBO,
 };
 
diff --git a/src/gallium/drivers/freedreno/a6xx/fd6_const.cc b/src/gallium/drivers/freedreno/a6xx/fd6_const.cc
index 4f2f80f7ac0..4dba336d99b 100644
--- a/src/gallium/drivers/freedreno/a6xx/fd6_const.cc
+++ b/src/gallium/drivers/freedreno/a6xx/fd6_const.cc
@@ -527,7 +527,7 @@ fd6_emit_cs_user_consts(struct fd_context *ctx,
                         struct fd_ringbuffer *ring,
                         struct fd6_compute_state *cs)
 {
-   emit_user_consts<CHIP>(cs->v, ring, &ctx->constbuf[PIPE_SHADER_COMPUTE]);
+   emit_user_consts<CHIP>(cs->v, ring, &ctx->constbuf[MESA_SHADER_COMPUTE]);
 }
 FD_GENX(fd6_emit_cs_user_consts);
 
diff --git a/src/gallium/drivers/freedreno/a6xx/fd6_context.cc b/src/gallium/drivers/freedreno/a6xx/fd6_context.cc
index 1b7d21f1de6..6cd91281459 100644
--- a/src/gallium/drivers/freedreno/a6xx/fd6_context.cc
+++ b/src/gallium/drivers/freedreno/a6xx/fd6_context.cc
@@ -187,7 +187,7 @@ setup_state_map(struct fd_context *ctx)
                              BIT(FD6_GROUP_GS_TEX));
    fd_context_add_shader_map(ctx, MESA_SHADER_FRAGMENT, FD_DIRTY_SHADER_TEX,
                              BIT(FD6_GROUP_FS_TEX));
-   fd_context_add_shader_map(ctx, PIPE_SHADER_COMPUTE, FD_DIRTY_SHADER_TEX,
+   fd_context_add_shader_map(ctx, MESA_SHADER_COMPUTE, FD_DIRTY_SHADER_TEX,
                              BIT(FD6_GROUP_CS_TEX));
 
    fd_context_add_shader_map(ctx, MESA_SHADER_VERTEX,
@@ -209,7 +209,7 @@ setup_state_map(struct fd_context *ctx)
    fd_context_add_shader_map(ctx, MESA_SHADER_FRAGMENT,
                              FD_DIRTY_SHADER_SSBO | FD_DIRTY_SHADER_IMAGE,
                              BIT(FD6_GROUP_FS_BINDLESS));
-   fd_context_add_shader_map(ctx, PIPE_SHADER_COMPUTE,
+   fd_context_add_shader_map(ctx, MESA_SHADER_COMPUTE,
                              FD_DIRTY_SHADER_SSBO | FD_DIRTY_SHADER_IMAGE,
                              BIT(FD6_GROUP_CS_BINDLESS));
    fd_context_add_shader_map(ctx, MESA_SHADER_FRAGMENT,
diff --git a/src/gallium/drivers/freedreno/a6xx/fd6_emit.cc b/src/gallium/drivers/freedreno/a6xx/fd6_emit.cc
index 46b2e48e75d..1d01cfef143 100644
--- a/src/gallium/drivers/freedreno/a6xx/fd6_emit.cc
+++ b/src/gallium/drivers/freedreno/a6xx/fd6_emit.cc
@@ -770,13 +770,13 @@ fd6_emit_cs_state(struct fd_context *ctx, struct fd_ringbuffer *ring,
       case FD6_GROUP_CS_TEX:
          fd6_state_take_group(
                &state,
-               tex_state(ctx, PIPE_SHADER_COMPUTE),
+               tex_state(ctx, MESA_SHADER_COMPUTE),
                FD6_GROUP_CS_TEX);
          break;
       case FD6_GROUP_CS_BINDLESS:
          fd6_state_take_group(
                &state,
-               fd6_build_bindless_state<CHIP>(ctx, PIPE_SHADER_COMPUTE, false),
+               fd6_build_bindless_state<CHIP>(ctx, MESA_SHADER_COMPUTE, false),
                FD6_GROUP_CS_BINDLESS);
          break;
       default:
diff --git a/src/gallium/drivers/freedreno/a6xx/fd6_image.cc b/src/gallium/drivers/freedreno/a6xx/fd6_image.cc
index 9de57b15ae1..ed068c48279 100644
--- a/src/gallium/drivers/freedreno/a6xx/fd6_image.cc
+++ b/src/gallium/drivers/freedreno/a6xx/fd6_image.cc
@@ -126,7 +126,7 @@ descriptor_set(struct fd_context *ctx, enum pipe_shader_type shader)
 {
    struct fd6_context *fd6_ctx = fd6_context(ctx);
 
-   if (shader == PIPE_SHADER_COMPUTE)
+   if (shader == MESA_SHADER_COMPUTE)
       return &fd6_ctx->cs_descriptor_set;
 
    unsigned idx = ir3_shader_descriptor_set(shader);
@@ -269,7 +269,7 @@ fd6_build_bindless_state(struct fd_context *ctx, enum pipe_shader_type shader,
 
    fd_ringbuffer_attach_bo(ring, set->bo);
 
-   if (shader == PIPE_SHADER_COMPUTE) {
+   if (shader == MESA_SHADER_COMPUTE) {
       OUT_REG(ring,
          SP_UPDATE_CNTL(
             CHIP,
diff --git a/src/gallium/drivers/freedreno/a6xx/fd6_texture.cc b/src/gallium/drivers/freedreno/a6xx/fd6_texture.cc
index 3393139d85d..0780496491a 100644
--- a/src/gallium/drivers/freedreno/a6xx/fd6_texture.cc
+++ b/src/gallium/drivers/freedreno/a6xx/fd6_texture.cc
@@ -645,7 +645,7 @@ build_texture_state(struct fd_context *ctx, enum pipe_shader_type type,
       tex_const_reg = REG_A6XX_SP_PS_TEXMEMOBJ_BASE;
       tex_count_reg = REG_A6XX_SP_PS_TSIZE;
       break;
-   case PIPE_SHADER_COMPUTE:
+   case MESA_SHADER_COMPUTE:
       sb = SB6_CS_TEX;
       opcode = CP_LOAD_STATE6_FRAG;
       tex_samp_reg = REG_A6XX_SP_CS_SAMPLER_BASE;
diff --git a/src/gallium/drivers/freedreno/freedreno_draw.c b/src/gallium/drivers/freedreno/freedreno_draw.c
index a0d11640321..cfa8af95393 100644
--- a/src/gallium/drivers/freedreno/freedreno_draw.c
+++ b/src/gallium/drivers/freedreno/freedreno_draw.c
@@ -558,7 +558,7 @@ fd_launch_grid(struct pipe_context *pctx,
 {
    struct fd_context *ctx = fd_context(pctx);
    const struct fd_shaderbuf_stateobj *so =
-      &ctx->shaderbuf[PIPE_SHADER_COMPUTE];
+      &ctx->shaderbuf[MESA_SHADER_COMPUTE];
    struct fd_batch *batch, *save_batch = NULL;
 
    if (!fd_render_condition_check(pctx))
@@ -577,8 +577,8 @@ fd_launch_grid(struct pipe_context *pctx,
    u_foreach_bit (i, so->enabled_mask & ~so->writable_mask)
       resource_read(batch, so->sb[i].buffer);
 
-   u_foreach_bit (i, ctx->shaderimg[PIPE_SHADER_COMPUTE].enabled_mask) {
-      struct pipe_image_view *img = &ctx->shaderimg[PIPE_SHADER_COMPUTE].si[i];
+   u_foreach_bit (i, ctx->shaderimg[MESA_SHADER_COMPUTE].enabled_mask) {
+      struct pipe_image_view *img = &ctx->shaderimg[MESA_SHADER_COMPUTE].si[i];
       if (img->access & PIPE_IMAGE_ACCESS_WRITE)
          resource_written(batch, img->resource);
       else
@@ -586,12 +586,12 @@ fd_launch_grid(struct pipe_context *pctx,
    }
 
    /* UBO's are read */
-   u_foreach_bit (i, ctx->constbuf[PIPE_SHADER_COMPUTE].enabled_mask)
-      resource_read(batch, ctx->constbuf[PIPE_SHADER_COMPUTE].cb[i].buffer);
+   u_foreach_bit (i, ctx->constbuf[MESA_SHADER_COMPUTE].enabled_mask)
+      resource_read(batch, ctx->constbuf[MESA_SHADER_COMPUTE].cb[i].buffer);
 
    /* Mark textures as being read */
-   u_foreach_bit (i, ctx->tex[PIPE_SHADER_COMPUTE].valid_textures)
-      resource_read(batch, ctx->tex[PIPE_SHADER_COMPUTE].textures[i]->texture);
+   u_foreach_bit (i, ctx->tex[MESA_SHADER_COMPUTE].valid_textures)
+      resource_read(batch, ctx->tex[MESA_SHADER_COMPUTE].textures[i]->texture);
 
    /* For global buffers, we don't really know if read or written, so assume
     * the worst:
diff --git a/src/gallium/drivers/freedreno/freedreno_screen.c b/src/gallium/drivers/freedreno/freedreno_screen.c
index 126df86a441..8467761b62c 100644
--- a/src/gallium/drivers/freedreno/freedreno_screen.c
+++ b/src/gallium/drivers/freedreno/freedreno_screen.c
@@ -203,7 +203,7 @@ fd_query_memory_info(struct pipe_screen *pscreen,
 static void
 fd_init_shader_caps(struct fd_screen *screen)
 {
-   for (unsigned i = 0; i <= PIPE_SHADER_COMPUTE; i++) {
+   for (unsigned i = 0; i <= MESA_SHADER_COMPUTE; i++) {
       struct pipe_shader_caps *caps =
          (struct pipe_shader_caps *)&screen->base.shader_caps[i];
 
@@ -216,7 +216,7 @@ fd_init_shader_caps(struct fd_screen *screen)
          if (screen->info->a6xx.is_a702)
             continue;
          break;
-      case PIPE_SHADER_COMPUTE:
+      case MESA_SHADER_COMPUTE:
          if (!has_compute(screen))
             continue;
          break;
@@ -261,7 +261,7 @@ fd_init_shader_caps(struct fd_screen *screen)
       caps->int16 =
       caps->fp16 =
          (is_a5xx(screen) || is_a6xx(screen)) &&
-         (i == PIPE_SHADER_COMPUTE || i == MESA_SHADER_FRAGMENT) &&
+         (i == MESA_SHADER_COMPUTE || i == MESA_SHADER_FRAGMENT) &&
          !FD_DBG(NOFP16);
       caps->glsl_16bit_load_dst = true;
 
@@ -273,7 +273,7 @@ fd_init_shader_caps(struct fd_screen *screen)
          /* tgsi_to_nir doesn't support all stages: */
          COND(i == MESA_SHADER_VERTEX ||
               i == MESA_SHADER_FRAGMENT ||
-              i == PIPE_SHADER_COMPUTE,
+              i == MESA_SHADER_COMPUTE,
               1 << PIPE_SHADER_IR_TGSI);
 
       if (is_a6xx(screen)) {
@@ -302,7 +302,7 @@ fd_init_shader_caps(struct fd_screen *screen)
           * but images also need texture state for read access
           * (isam/isam.3d)
           */
-         if (i == MESA_SHADER_FRAGMENT || i == PIPE_SHADER_COMPUTE) {
+         if (i == MESA_SHADER_FRAGMENT || i == MESA_SHADER_COMPUTE) {
             caps->max_shader_buffers =
             caps->max_shader_images = 24;
          }
diff --git a/src/gallium/drivers/freedreno/freedreno_state.c b/src/gallium/drivers/freedreno/freedreno_state.c
index def7f4eea1f..38d99c8b3a9 100644
--- a/src/gallium/drivers/freedreno/freedreno_state.c
+++ b/src/gallium/drivers/freedreno/freedreno_state.c
@@ -714,7 +714,7 @@ fd_bind_compute_state(struct pipe_context *pctx, void *state) in_dt
 {
    struct fd_context *ctx = fd_context(pctx);
    ctx->compute = state;
-   fd_context_dirty_shader(ctx, PIPE_SHADER_COMPUTE, FD_DIRTY_SHADER_PROG);
+   fd_context_dirty_shader(ctx, MESA_SHADER_COMPUTE, FD_DIRTY_SHADER_PROG);
 }
 
 /* used by clover to bind global objects, returning the bo address
diff --git a/src/gallium/drivers/freedreno/ir3/ir3_const.h b/src/gallium/drivers/freedreno/ir3/ir3_const.h
index 80f04808602..395a505815d 100644
--- a/src/gallium/drivers/freedreno/ir3/ir3_const.h
+++ b/src/gallium/drivers/freedreno/ir3/ir3_const.h
@@ -672,7 +672,7 @@ ir3_emit_cs_consts(const struct ir3_shader_variant *v,
 {
    assert(gl_shader_stage_is_compute(v->type));
 
-   emit_common_consts(v, ring, ctx, PIPE_SHADER_COMPUTE);
+   emit_common_consts(v, ring, ctx, MESA_SHADER_COMPUTE);
 
    ir3_emit_cs_driver_params(v, ring, ctx, info);
 }
diff --git a/src/gallium/drivers/freedreno/ir3/ir3_descriptor.h b/src/gallium/drivers/freedreno/ir3/ir3_descriptor.h
index 5b38384f2bb..5c80437e3ce 100644
--- a/src/gallium/drivers/freedreno/ir3/ir3_descriptor.h
+++ b/src/gallium/drivers/freedreno/ir3/ir3_descriptor.h
@@ -40,7 +40,7 @@ ir3_shader_descriptor_set(enum pipe_shader_type shader)
    case MESA_SHADER_TESS_EVAL: return 2;
    case MESA_SHADER_GEOMETRY:  return 3;
    case MESA_SHADER_FRAGMENT:  return 4;
-   case PIPE_SHADER_COMPUTE:   return 0;
+   case MESA_SHADER_COMPUTE:   return 0;
    case MESA_SHADER_KERNEL:    return 0;
    default:
       UNREACHABLE("bad shader stage");
diff --git a/src/gallium/drivers/iris/iris_screen.c b/src/gallium/drivers/iris/iris_screen.c
index d501a3ce15f..69ad2de52d2 100644
--- a/src/gallium/drivers/iris/iris_screen.c
+++ b/src/gallium/drivers/iris/iris_screen.c
@@ -201,7 +201,7 @@ iris_get_video_memory(struct iris_screen *screen)
 static void
 iris_init_shader_caps(struct iris_screen *screen)
 {
-   for (unsigned i = 0; i <= PIPE_SHADER_COMPUTE; i++) {
+   for (unsigned i = 0; i <= MESA_SHADER_COMPUTE; i++) {
       struct pipe_shader_caps *caps =
          (struct pipe_shader_caps *)&screen->base.shader_caps[i];
 
diff --git a/src/gallium/drivers/llvmpipe/lp_screen.c b/src/gallium/drivers/llvmpipe/lp_screen.c
index c36a6866ee0..48e47b993bc 100644
--- a/src/gallium/drivers/llvmpipe/lp_screen.c
+++ b/src/gallium/drivers/llvmpipe/lp_screen.c
@@ -128,7 +128,7 @@ llvmpipe_init_shader_caps(struct pipe_screen *screen)
 
       switch (i) {
       case MESA_SHADER_FRAGMENT:
-      case PIPE_SHADER_COMPUTE:
+      case MESA_SHADER_COMPUTE:
       case PIPE_SHADER_MESH:
       case PIPE_SHADER_TASK:
          gallivm_init_shader_caps(caps);
diff --git a/src/gallium/drivers/llvmpipe/lp_state_cs.c b/src/gallium/drivers/llvmpipe/lp_state_cs.c
index 2f9ef5b7da2..bacbd5120a6 100644
--- a/src/gallium/drivers/llvmpipe/lp_state_cs.c
+++ b/src/gallium/drivers/llvmpipe/lp_state_cs.c
@@ -1457,7 +1457,7 @@ static void
 llvmpipe_update_cs(struct llvmpipe_context *lp)
 {
    struct lp_compute_shader_variant *variant;
-   variant = llvmpipe_update_cs_variant(lp, PIPE_SHADER_COMPUTE, lp->cs);
+   variant = llvmpipe_update_cs_variant(lp, MESA_SHADER_COMPUTE, lp->cs);
    /* Bind this variant */
    lp_cs_ctx_set_cs_variant(lp->csctx, variant);
 }
@@ -1643,32 +1643,32 @@ llvmpipe_cs_update_derived(struct llvmpipe_context *llvmpipe)
 {
    if (llvmpipe->cs_dirty & LP_CSNEW_CONSTANTS) {
       lp_csctx_set_cs_constants(llvmpipe->csctx,
-                                ARRAY_SIZE(llvmpipe->constants[PIPE_SHADER_COMPUTE]),
-                                llvmpipe->constants[PIPE_SHADER_COMPUTE]);
+                                ARRAY_SIZE(llvmpipe->constants[MESA_SHADER_COMPUTE]),
+                                llvmpipe->constants[MESA_SHADER_COMPUTE]);
       update_csctx_consts(llvmpipe, llvmpipe->csctx);
    }
 
    if (llvmpipe->cs_dirty & LP_CSNEW_SSBOS) {
       lp_csctx_set_cs_ssbos(llvmpipe->csctx,
-                            ARRAY_SIZE(llvmpipe->ssbos[PIPE_SHADER_COMPUTE]),
-                            llvmpipe->ssbos[PIPE_SHADER_COMPUTE]);
+                            ARRAY_SIZE(llvmpipe->ssbos[MESA_SHADER_COMPUTE]),
+                            llvmpipe->ssbos[MESA_SHADER_COMPUTE]);
       update_csctx_ssbo(llvmpipe, llvmpipe->csctx);
    }
 
    if (llvmpipe->cs_dirty & LP_CSNEW_SAMPLER_VIEW)
       lp_csctx_set_sampler_views(llvmpipe->csctx,
-                                 llvmpipe->num_sampler_views[PIPE_SHADER_COMPUTE],
-                                 llvmpipe->sampler_views[PIPE_SHADER_COMPUTE]);
+                                 llvmpipe->num_sampler_views[MESA_SHADER_COMPUTE],
+                                 llvmpipe->sampler_views[MESA_SHADER_COMPUTE]);
 
    if (llvmpipe->cs_dirty & LP_CSNEW_SAMPLER)
       lp_csctx_set_sampler_state(llvmpipe->csctx,
-                                 llvmpipe->num_samplers[PIPE_SHADER_COMPUTE],
-                                 llvmpipe->samplers[PIPE_SHADER_COMPUTE]);
+                                 llvmpipe->num_samplers[MESA_SHADER_COMPUTE],
+                                 llvmpipe->samplers[MESA_SHADER_COMPUTE]);
 
    if (llvmpipe->cs_dirty & LP_CSNEW_IMAGES)
       lp_csctx_set_cs_images(llvmpipe->csctx,
-                              ARRAY_SIZE(llvmpipe->images[PIPE_SHADER_COMPUTE]),
-                              llvmpipe->images[PIPE_SHADER_COMPUTE]);
+                              ARRAY_SIZE(llvmpipe->images[MESA_SHADER_COMPUTE]),
+                              llvmpipe->images[MESA_SHADER_COMPUTE]);
 
    if (llvmpipe->cs_dirty & (LP_CSNEW_CS |
                              LP_CSNEW_IMAGES |
diff --git a/src/gallium/drivers/llvmpipe/lp_state_fs.c b/src/gallium/drivers/llvmpipe/lp_state_fs.c
index e9f54773684..1790c1dfeb0 100644
--- a/src/gallium/drivers/llvmpipe/lp_state_fs.c
+++ b/src/gallium/drivers/llvmpipe/lp_state_fs.c
@@ -4260,7 +4260,7 @@ llvmpipe_set_constant_buffer(struct pipe_context *pipe,
                                       index, data, size);
       break;
    }
-   case PIPE_SHADER_COMPUTE:
+   case MESA_SHADER_COMPUTE:
       llvmpipe->cs_dirty |= LP_CSNEW_CONSTANTS;
       break;
    case MESA_SHADER_FRAGMENT:
@@ -4315,7 +4315,7 @@ llvmpipe_set_shader_buffers(struct pipe_context *pipe,
                                        i, data, size);
          break;
       }
-      case PIPE_SHADER_COMPUTE:
+      case MESA_SHADER_COMPUTE:
          llvmpipe->cs_dirty |= LP_CSNEW_SSBOS;
          break;
       case PIPE_SHADER_TASK:
@@ -4368,7 +4368,7 @@ llvmpipe_set_shader_images(struct pipe_context *pipe,
       draw_set_images(llvmpipe->draw, shader, llvmpipe->images[shader],
                       start_slot + count);
       break;
-   case PIPE_SHADER_COMPUTE:
+   case MESA_SHADER_COMPUTE:
       llvmpipe->cs_dirty |= LP_CSNEW_IMAGES;
       break;
    case MESA_SHADER_FRAGMENT:
diff --git a/src/gallium/drivers/llvmpipe/lp_state_sampler.c b/src/gallium/drivers/llvmpipe/lp_state_sampler.c
index 692bfce2dab..0b97ba419be 100644
--- a/src/gallium/drivers/llvmpipe/lp_state_sampler.c
+++ b/src/gallium/drivers/llvmpipe/lp_state_sampler.c
@@ -106,7 +106,7 @@ llvmpipe_bind_sampler_states(struct pipe_context *pipe,
                         llvmpipe->samplers[shader],
                         llvmpipe->num_samplers[shader]);
       break;
-   case PIPE_SHADER_COMPUTE:
+   case MESA_SHADER_COMPUTE:
       llvmpipe->cs_dirty |= LP_CSNEW_SAMPLER;
       break;
    case MESA_SHADER_FRAGMENT:
@@ -189,7 +189,7 @@ llvmpipe_set_sampler_views(struct pipe_context *pipe,
                              llvmpipe->sampler_views[shader],
                              llvmpipe->num_sampler_views[shader]);
       break;
-   case PIPE_SHADER_COMPUTE:
+   case MESA_SHADER_COMPUTE:
       llvmpipe->cs_dirty |= LP_CSNEW_SAMPLER_VIEW;
       break;
    case MESA_SHADER_FRAGMENT:
diff --git a/src/gallium/drivers/nouveau/codegen/nv50_ir.cpp b/src/gallium/drivers/nouveau/codegen/nv50_ir.cpp
index 7a51ec5354d..9e7e32cc052 100644
--- a/src/gallium/drivers/nouveau/codegen/nv50_ir.cpp
+++ b/src/gallium/drivers/nouveau/codegen/nv50_ir.cpp
@@ -1250,7 +1250,7 @@ nv50_ir_init_prog_info(struct nv50_ir_prog_info *info,
       info_out->prop.gp.instanceCount = 1;
       info_out->prop.gp.maxVertices = 1;
    }
-   if (info->type == PIPE_SHADER_COMPUTE) {
+   if (info->type == MESA_SHADER_COMPUTE) {
       info->prop.cp.numThreads[0] =
       info->prop.cp.numThreads[1] =
       info->prop.cp.numThreads[2] = 1;
diff --git a/src/gallium/drivers/nouveau/codegen/nv50_ir_serialize.cpp b/src/gallium/drivers/nouveau/codegen/nv50_ir_serialize.cpp
index f3fb47c84c6..8f685575ec3 100644
--- a/src/gallium/drivers/nouveau/codegen/nv50_ir_serialize.cpp
+++ b/src/gallium/drivers/nouveau/codegen/nv50_ir_serialize.cpp
@@ -29,7 +29,7 @@ nv50_ir_prog_info_serialize(struct blob *blob, struct nv50_ir_prog_info *info)
 
    nir_serialize(blob, info->bin.nir, true);
 
-   if (info->type == PIPE_SHADER_COMPUTE)
+   if (info->type == MESA_SHADER_COMPUTE)
       blob_write_bytes(blob, &info->prop.cp, sizeof(info->prop.cp));
 
    blob_write_bytes(blob, &info->io, sizeof(info->io));
@@ -122,7 +122,7 @@ nv50_ir_prog_info_out_serialize(struct blob *blob,
       case MESA_SHADER_FRAGMENT:
          blob_write_bytes(blob, &info_out->prop.fp, sizeof(info_out->prop.fp));
          break;
-      case PIPE_SHADER_COMPUTE:
+      case MESA_SHADER_COMPUTE:
          blob_write_bytes(blob, &info_out->prop.cp, sizeof(info_out->prop.cp));
          break;
       default:
@@ -242,7 +242,7 @@ nv50_ir_prog_info_out_deserialize(void *data, size_t size, size_t offset,
       case MESA_SHADER_FRAGMENT:
          blob_copy_bytes(&reader, &info_out->prop.fp, sizeof(info_out->prop.fp));
          break;
-      case PIPE_SHADER_COMPUTE:
+      case MESA_SHADER_COMPUTE:
          blob_copy_bytes(&reader, &info_out->prop.cp, sizeof(info_out->prop.cp));
          break;
       default:
diff --git a/src/gallium/drivers/nouveau/codegen/nv50_ir_target.h b/src/gallium/drivers/nouveau/codegen/nv50_ir_target.h
index 691479c5c93..aeabfaff877 100644
--- a/src/gallium/drivers/nouveau/codegen/nv50_ir_target.h
+++ b/src/gallium/drivers/nouveau/codegen/nv50_ir_target.h
@@ -176,7 +176,7 @@ public:
 
    virtual void parseDriverInfo(const struct nv50_ir_prog_info *info,
                                 const struct nv50_ir_prog_info_out *info_out) {
-      if (info_out->type == PIPE_SHADER_COMPUTE) {
+      if (info_out->type == MESA_SHADER_COMPUTE) {
          threads = info->prop.cp.numThreads[0] *
             info->prop.cp.numThreads[1] *
             info->prop.cp.numThreads[2];
diff --git a/src/gallium/drivers/nouveau/nv50/nv50_context.h b/src/gallium/drivers/nouveau/nv50/nv50_context.h
index eb9d7a14b44..3415aca84dd 100644
--- a/src/gallium/drivers/nouveau/nv50/nv50_context.h
+++ b/src/gallium/drivers/nouveau/nv50/nv50_context.h
@@ -250,7 +250,7 @@ nv50_context_shader_stage(unsigned pipe)
    case MESA_SHADER_VERTEX: return NV50_SHADER_STAGE_VERTEX;
    case MESA_SHADER_FRAGMENT: return NV50_SHADER_STAGE_FRAGMENT;
    case MESA_SHADER_GEOMETRY: return NV50_SHADER_STAGE_GEOMETRY;
-   case PIPE_SHADER_COMPUTE: return NV50_SHADER_STAGE_COMPUTE;
+   case MESA_SHADER_COMPUTE: return NV50_SHADER_STAGE_COMPUTE;
    default:
       assert(!"invalid/unhandled shader type");
       return 0;
diff --git a/src/gallium/drivers/nouveau/nv50/nv50_program.c b/src/gallium/drivers/nouveau/nv50/nv50_program.c
index 16786c79356..45a24189ce9 100644
--- a/src/gallium/drivers/nouveau/nv50/nv50_program.c
+++ b/src/gallium/drivers/nouveau/nv50/nv50_program.c
@@ -262,7 +262,7 @@ nv50_program_assign_varying_slots(struct nv50_ir_prog_info_out *info)
       return nv50_vertprog_assign_slots(info);
    case MESA_SHADER_FRAGMENT:
       return nv50_fragprog_assign_slots(info);
-   case PIPE_SHADER_COMPUTE:
+   case MESA_SHADER_COMPUTE:
       return 0;
    default:
       return -1;
@@ -370,7 +370,7 @@ nv50_program_translate(struct nv50_program *prog, uint16_t chipset,
    prog->gp.has_layer = 0;
    prog->gp.has_viewport = 0;
 
-   if (prog->type == PIPE_SHADER_COMPUTE)
+   if (prog->type == MESA_SHADER_COMPUTE)
       info->prop.cp.inputOffset = 0x14;
 
    info_out.driverPriv = prog;
@@ -430,7 +430,7 @@ nv50_program_translate(struct nv50_program *prog, uint16_t chipset,
       }
       prog->gp.vert_count = CLAMP(info_out.prop.gp.maxVertices, 1, 1024);
    } else
-   if (prog->type == PIPE_SHADER_COMPUTE) {
+   if (prog->type == MESA_SHADER_COMPUTE) {
       for (i = 0; i < NV50_MAX_GLOBALS; i++) {
          prog->cp.gmem[i] = (struct nv50_gmem_state){
             .valid = info_out.prop.cp.gmem[i].valid,
@@ -468,7 +468,7 @@ nv50_program_upload_code(struct nv50_context *nv50, struct nv50_program *prog)
    case MESA_SHADER_VERTEX:   heap = nv50->screen->vp_code_heap; break;
    case MESA_SHADER_GEOMETRY: heap = nv50->screen->gp_code_heap; break;
    case MESA_SHADER_FRAGMENT: heap = nv50->screen->fp_code_heap; break;
-   case PIPE_SHADER_COMPUTE:  heap = nv50->screen->fp_code_heap; break;
+   case MESA_SHADER_COMPUTE:  heap = nv50->screen->fp_code_heap; break;
    default:
       assert(!"invalid program type");
       return false;
@@ -493,7 +493,7 @@ nv50_program_upload_code(struct nv50_context *nv50, struct nv50_program *prog)
       }
    }
 
-   if (prog->type == PIPE_SHADER_COMPUTE) {
+   if (prog->type == MESA_SHADER_COMPUTE) {
       /* CP code must be uploaded in FP code segment. */
       prog_type = NV50_SHADER_STAGE_FRAGMENT;
    } else {
diff --git a/src/gallium/drivers/nouveau/nv50/nv50_query_hw_sm.c b/src/gallium/drivers/nouveau/nv50/nv50_query_hw_sm.c
index bb296e63143..c1f78a0677e 100644
--- a/src/gallium/drivers/nouveau/nv50/nv50_query_hw_sm.c
+++ b/src/gallium/drivers/nouveau/nv50/nv50_query_hw_sm.c
@@ -228,7 +228,7 @@ nv50_hw_sm_end_query(struct nv50_context *nv50, struct nv50_hw_query *hq)
 
    if (unlikely(!screen->pm.prog)) {
       struct nv50_program *prog = CALLOC_STRUCT(nv50_program);
-      prog->type = PIPE_SHADER_COMPUTE;
+      prog->type = MESA_SHADER_COMPUTE;
       prog->translated = true;
       prog->max_gpr = 7;
       prog->code = (uint32_t *)nv50_read_hw_sm_counters_code;
diff --git a/src/gallium/drivers/nouveau/nv50/nv50_screen.c b/src/gallium/drivers/nouveau/nv50/nv50_screen.c
index 1d7d18174be..7fbf86f9222 100644
--- a/src/gallium/drivers/nouveau/nv50/nv50_screen.c
+++ b/src/gallium/drivers/nouveau/nv50/nv50_screen.c
@@ -104,7 +104,7 @@ nv50_screen_is_format_supported(struct pipe_screen *pscreen,
 static void
 nv50_init_shader_caps(struct nv50_screen *screen)
 {
-   for (unsigned i = 0; i <= PIPE_SHADER_COMPUTE; i++) {
+   for (unsigned i = 0; i <= MESA_SHADER_COMPUTE; i++) {
       struct pipe_shader_caps *caps =
          (struct pipe_shader_caps *)&screen->base.base.shader_caps[i];
 
@@ -130,7 +130,7 @@ nv50_init_shader_caps(struct nv50_screen *screen)
       /* The chip could handle more sampler views than samplers */
       caps->max_sampler_views = MIN2(16, PIPE_MAX_SAMPLERS);
       caps->max_shader_buffers =
-      caps->max_shader_images = i == PIPE_SHADER_COMPUTE ? NV50_MAX_GLOBALS - 1 : 0;
+      caps->max_shader_images = i == MESA_SHADER_COMPUTE ? NV50_MAX_GLOBALS - 1 : 0;
       caps->supported_irs = 1 << PIPE_SHADER_IR_NIR;
    }
 }
diff --git a/src/gallium/drivers/nouveau/nv50/nv50_state.c b/src/gallium/drivers/nouveau/nv50/nv50_state.c
index 94a461eda17..87de8c4fa47 100644
--- a/src/gallium/drivers/nouveau/nv50/nv50_state.c
+++ b/src/gallium/drivers/nouveau/nv50/nv50_state.c
@@ -835,7 +835,7 @@ nv50_cp_state_create(struct pipe_context *pipe,
    prog = CALLOC_STRUCT(nv50_program);
    if (!prog)
       return NULL;
-   prog->type = PIPE_SHADER_COMPUTE;
+   prog->type = MESA_SHADER_COMPUTE;
 
    switch(cso->ir_type) {
    case PIPE_SHADER_IR_TGSI: {
@@ -893,7 +893,7 @@ nv50_set_constant_buffer(struct pipe_context *pipe,
    const unsigned s = nv50_context_shader_stage(shader);
    const unsigned i = index;
 
-   if (unlikely(shader == PIPE_SHADER_COMPUTE)) {
+   if (unlikely(shader == MESA_SHADER_COMPUTE)) {
       if (nv50->constbuf[s][i].user)
          nv50->constbuf[s][i].u.buf = NULL;
       else
diff --git a/src/gallium/drivers/nouveau/nvc0/nvc0_context.h b/src/gallium/drivers/nouveau/nvc0/nvc0_context.h
index 64e6a9abb36..7e75363cd5d 100644
--- a/src/gallium/drivers/nouveau/nvc0/nvc0_context.h
+++ b/src/gallium/drivers/nouveau/nvc0/nvc0_context.h
@@ -301,7 +301,7 @@ nvc0_shader_stage(unsigned pipe)
    case MESA_SHADER_TESS_EVAL: return 2;
    case MESA_SHADER_GEOMETRY: return 3;
    case MESA_SHADER_FRAGMENT: return 4;
-   case PIPE_SHADER_COMPUTE: return 5;
+   case MESA_SHADER_COMPUTE: return 5;
    default:
       assert(!"invalid PIPE_SHADER type");
       return 0;
diff --git a/src/gallium/drivers/nouveau/nvc0/nvc0_program.c b/src/gallium/drivers/nouveau/nvc0/nvc0_program.c
index 9e969986cc4..8851503ccb9 100644
--- a/src/gallium/drivers/nouveau/nvc0/nvc0_program.c
+++ b/src/gallium/drivers/nouveau/nvc0/nvc0_program.c
@@ -555,7 +555,7 @@ nvc0_program_dump(struct nvc0_program *prog)
 {
    unsigned pos;
 
-   if (prog->type != PIPE_SHADER_COMPUTE) {
+   if (prog->type != MESA_SHADER_COMPUTE) {
       _debug_printf("dumping HDR for type %i\n", prog->type);
       for (pos = 0; pos < ARRAY_SIZE(prog->hdr); ++pos)
          _debug_printf("HDR[%02"PRIxPTR"] = 0x%08x\n",
@@ -618,7 +618,7 @@ nvc0_program_translate(struct nvc0_program *prog, uint16_t chipset,
       info->io.bindlessBase = NVC0_CB_AUX_BINDLESS_INFO(0);
    }
 
-   if (prog->type == PIPE_SHADER_COMPUTE) {
+   if (prog->type == MESA_SHADER_COMPUTE) {
       if (info->target >= NVISA_GK104_CHIPSET) {
          info->io.auxCBSlot = 7;
          info->io.msInfoCBSlot = 7;
@@ -705,7 +705,7 @@ nvc0_program_translate(struct nvc0_program *prog, uint16_t chipset,
    case MESA_SHADER_FRAGMENT:
       ret = nvc0_fp_gen_header(prog, &info_out);
       break;
-   case PIPE_SHADER_COMPUTE:
+   case MESA_SHADER_COMPUTE:
       break;
    default:
       ret = -1;
@@ -762,7 +762,7 @@ static inline int
 nvc0_program_alloc_code(struct nvc0_context *nvc0, struct nvc0_program *prog)
 {
    struct nvc0_screen *screen = nvc0->screen;
-   const bool is_cp = prog->type == PIPE_SHADER_COMPUTE;
+   const bool is_cp = prog->type == MESA_SHADER_COMPUTE;
    int ret;
    uint32_t size = prog->code_size;
 
@@ -814,7 +814,7 @@ static inline void
 nvc0_program_upload_code(struct nvc0_context *nvc0, struct nvc0_program *prog)
 {
    struct nvc0_screen *screen = nvc0->screen;
-   const bool is_cp = prog->type == PIPE_SHADER_COMPUTE;
+   const bool is_cp = prog->type == MESA_SHADER_COMPUTE;
    uint32_t code_pos = prog->code_base;
    uint32_t size_sph = 0;
 
@@ -862,7 +862,7 @@ bool
 nvc0_program_upload(struct nvc0_context *nvc0, struct nvc0_program *prog)
 {
    struct nvc0_screen *screen = nvc0->screen;
-   const bool is_cp = prog->type == PIPE_SHADER_COMPUTE;
+   const bool is_cp = prog->type == MESA_SHADER_COMPUTE;
    int ret;
    uint32_t size = prog->code_size;
 
@@ -923,7 +923,7 @@ nvc0_program_upload(struct nvc0_context *nvc0, struct nvc0_program *prog)
          }
          nvc0_program_upload_code(nvc0, progs[i]);
 
-         if (progs[i]->type == PIPE_SHADER_COMPUTE) {
+         if (progs[i]->type == MESA_SHADER_COMPUTE) {
             /* Caches have to be invalidated but the CP_START_ID will be
              * updated in the launch_grid functions. */
             BEGIN_NVC0(nvc0->base.pushbuf, NVC0_CP(FLUSH), 1);
diff --git a/src/gallium/drivers/nouveau/nvc0/nvc0_query_hw_sm.c b/src/gallium/drivers/nouveau/nvc0/nvc0_query_hw_sm.c
index 182f0d42f5b..86e63bab897 100644
--- a/src/gallium/drivers/nouveau/nvc0/nvc0_query_hw_sm.c
+++ b/src/gallium/drivers/nouveau/nvc0/nvc0_query_hw_sm.c
@@ -2470,7 +2470,7 @@ nvc0_hw_sm_get_program(struct nvc0_screen *screen)
    if (!prog)
       return NULL;
 
-   prog->type = PIPE_SHADER_COMPUTE;
+   prog->type = MESA_SHADER_COMPUTE;
    prog->translated = true;
 
    if (screen->base.class_3d >= GM107_3D_CLASS) {
diff --git a/src/gallium/drivers/nouveau/nvc0/nvc0_screen.c b/src/gallium/drivers/nouveau/nvc0/nvc0_screen.c
index 21aa5323b56..ed6b8041a42 100644
--- a/src/gallium/drivers/nouveau/nvc0/nvc0_screen.c
+++ b/src/gallium/drivers/nouveau/nvc0/nvc0_screen.c
@@ -116,7 +116,7 @@ nvc0_init_shader_caps(struct nvc0_screen *screen)
 {
    const uint16_t class_3d = screen->base.class_3d;
 
-   for (unsigned i = 0; i <= PIPE_SHADER_COMPUTE; i++) {
+   for (unsigned i = 0; i <= MESA_SHADER_COMPUTE; i++) {
       struct pipe_shader_caps *caps =
          (struct pipe_shader_caps *)&screen->base.base.shader_caps[i];
 
@@ -143,7 +143,7 @@ nvc0_init_shader_caps(struct nvc0_screen *screen)
       caps->max_shader_images =
          class_3d >= NVE4_3D_CLASS ||
          i == MESA_SHADER_FRAGMENT ||
-         i == PIPE_SHADER_COMPUTE ? NVC0_MAX_IMAGES : 0;
+         i == MESA_SHADER_COMPUTE ? NVC0_MAX_IMAGES : 0;
    }
 }
 
diff --git a/src/gallium/drivers/nouveau/nvc0/nvc0_state.c b/src/gallium/drivers/nouveau/nvc0/nvc0_state.c
index c2ab4b85662..a75dde41d62 100644
--- a/src/gallium/drivers/nouveau/nvc0/nvc0_state.c
+++ b/src/gallium/drivers/nouveau/nvc0/nvc0_state.c
@@ -724,7 +724,7 @@ nvc0_cp_state_create(struct pipe_context *pipe,
    prog = CALLOC_STRUCT(nvc0_program);
    if (!prog)
       return NULL;
-   prog->type = PIPE_SHADER_COMPUTE;
+   prog->type = MESA_SHADER_COMPUTE;
 
    prog->cp.smem_size = cso->static_shared_mem;
 
@@ -797,7 +797,7 @@ nvc0_set_constant_buffer(struct pipe_context *pipe,
    const unsigned s = nvc0_shader_stage(shader);
    const unsigned i = index;
 
-   if (unlikely(shader == PIPE_SHADER_COMPUTE)) {
+   if (unlikely(shader == MESA_SHADER_COMPUTE)) {
       if (nvc0->constbuf[s][i].user)
          nvc0->constbuf[s][i].u.buf = NULL;
       else
diff --git a/src/gallium/drivers/nouveau/nvc0/nve4_compute.c b/src/gallium/drivers/nouveau/nvc0/nve4_compute.c
index 042f661e7b5..f9ad79b81ef 100644
--- a/src/gallium/drivers/nouveau/nvc0/nve4_compute.c
+++ b/src/gallium/drivers/nouveau/nvc0/nve4_compute.c
@@ -315,7 +315,7 @@ nve4_compute_set_tex_handles(struct nvc0_context *nvc0)
    struct nouveau_pushbuf *push = nvc0->base.pushbuf;
    struct nvc0_screen *screen = nvc0->screen;
    uint64_t address;
-   const unsigned s = nvc0_shader_stage(PIPE_SHADER_COMPUTE);
+   const unsigned s = nvc0_shader_stage(MESA_SHADER_COMPUTE);
    unsigned i, n;
    uint32_t dirty = nvc0->textures_dirty[s] | nvc0->samplers_dirty[s];
 
diff --git a/src/gallium/drivers/panfrost/pan_cmdstream.c b/src/gallium/drivers/panfrost/pan_cmdstream.c
index 3e6a5f8332c..837d1a0ee15 100644
--- a/src/gallium/drivers/panfrost/pan_cmdstream.c
+++ b/src/gallium/drivers/panfrost/pan_cmdstream.c
@@ -1615,7 +1615,7 @@ panfrost_emit_shared_memory(struct panfrost_batch *batch,
 {
    struct panfrost_context *ctx = batch->ctx;
    struct panfrost_device *dev = pan_device(ctx->base.screen);
-   struct panfrost_compiled_shader *ss = ctx->prog[PIPE_SHADER_COMPUTE];
+   struct panfrost_compiled_shader *ss = ctx->prog[MESA_SHADER_COMPUTE];
    struct pan_ptr t = pan_pool_alloc_desc(&batch->pool.base, LOCAL_STORAGE);
 
    struct pan_compute_dim local_size = {grid->block[0], grid->block[1],
@@ -3501,7 +3501,7 @@ panfrost_launch_grid_on_batch(struct pipe_context *pipe,
          continue;
 
       struct panfrost_resource *buffer = pan_resource(*res);
-      panfrost_batch_write_rsrc(batch, buffer, PIPE_SHADER_COMPUTE);
+      panfrost_batch_write_rsrc(batch, buffer, MESA_SHADER_COMPUTE);
    }
 
    if (info->indirect && !PAN_GPU_SUPPORTS_DISPATCH_INDIRECT) {
@@ -3528,7 +3528,7 @@ panfrost_launch_grid_on_batch(struct pipe_context *pipe,
    /* Conservatively assume workgroup size changes every launch */
    ctx->dirty |= PAN_DIRTY_PARAMS;
 
-   panfrost_update_shader_state(batch, PIPE_SHADER_COMPUTE);
+   panfrost_update_shader_state(batch, MESA_SHADER_COMPUTE);
 
    /* We want our compute thread descriptor to be per job.
     * Save the global one, and restore it when we're done emitting
@@ -3539,7 +3539,7 @@ panfrost_launch_grid_on_batch(struct pipe_context *pipe,
 
    /* if indirect, mark the indirect buffer as being read */
    if (info->indirect)
-      panfrost_batch_read_rsrc(batch, pan_resource(info->indirect), PIPE_SHADER_COMPUTE);
+      panfrost_batch_read_rsrc(batch, pan_resource(info->indirect), MESA_SHADER_COMPUTE);
 
    /* launch it */
    JOBX(launch_grid)(batch, info);
@@ -3583,17 +3583,17 @@ panfrost_launch_afbc_conv_shader(struct panfrost_batch *batch, void *cso,
    };
 
    struct panfrost_constant_buffer *pbuf =
-      &batch->ctx->constant_buffer[PIPE_SHADER_COMPUTE];
-   saved_cso = batch->ctx->uncompiled[PIPE_SHADER_COMPUTE];
+      &batch->ctx->constant_buffer[MESA_SHADER_COMPUTE];
+   saved_cso = batch->ctx->uncompiled[MESA_SHADER_COMPUTE];
    util_copy_constant_buffer(&pbuf->cb[0], &saved_const, true);
 
    pctx->bind_compute_state(pctx, cso);
-   pctx->set_constant_buffer(pctx, PIPE_SHADER_COMPUTE, 0, false, cbuf);
+   pctx->set_constant_buffer(pctx, MESA_SHADER_COMPUTE, 0, false, cbuf);
 
    panfrost_launch_grid_on_batch(pctx, batch, &grid);
 
    pctx->bind_compute_state(pctx, saved_cso);
-   pctx->set_constant_buffer(pctx, PIPE_SHADER_COMPUTE, 0, true, &saved_const);
+   pctx->set_constant_buffer(pctx, MESA_SHADER_COMPUTE, 0, true, &saved_const);
 }
 
 #define LAUNCH_AFBC_CONV_SHADER(name, batch, rsrc, consts, nr_blocks)          \
@@ -3625,8 +3625,8 @@ panfrost_afbc_size(struct panfrost_batch *batch, struct panfrost_resource *src,
                      src->image.props.modifier,
                      u_minify(src->image.props.extent_px.height, level));
 
-   panfrost_batch_read_rsrc(batch, src, PIPE_SHADER_COMPUTE);
-   panfrost_batch_write_bo(batch, layout, PIPE_SHADER_COMPUTE);
+   panfrost_batch_read_rsrc(batch, src, MESA_SHADER_COMPUTE);
+   panfrost_batch_write_bo(batch, layout, MESA_SHADER_COMPUTE);
 
    LAUNCH_AFBC_CONV_SHADER(size, batch, src, consts, nr_sblocks);
 }
@@ -3661,9 +3661,9 @@ panfrost_afbc_pack(struct panfrost_batch *batch, struct panfrost_resource *src,
       .dst_stride = dst_stride_sb,
    };
 
-   panfrost_batch_read_rsrc(batch, src, PIPE_SHADER_COMPUTE);
-   panfrost_batch_write_bo(batch, dst, PIPE_SHADER_COMPUTE);
-   panfrost_batch_add_bo(batch, layout, PIPE_SHADER_COMPUTE);
+   panfrost_batch_read_rsrc(batch, src, MESA_SHADER_COMPUTE);
+   panfrost_batch_write_bo(batch, dst, MESA_SHADER_COMPUTE);
+   panfrost_batch_add_bo(batch, layout, MESA_SHADER_COMPUTE);
 
    LAUNCH_AFBC_CONV_SHADER(pack, batch, src, consts, nr_sblocks);
 }
@@ -3766,7 +3766,7 @@ panfrost_mtk_detile_compute(struct panfrost_context *ctx, struct pipe_blit_info
    panfrost_flush_all_batches(ctx, "mtk_detile pre-barrier");
 
    struct panfrost_batch *batch = panfrost_get_batch_for_fbo(ctx);
-   pipe->set_shader_images(pipe, PIPE_SHADER_COMPUTE, 0, 4, 0, image);
+   pipe->set_shader_images(pipe, MESA_SHADER_COMPUTE, 0, 4, 0, image);
 
    /* launch the compute shader */
    struct pan_mod_convert_shader_data *shader =
@@ -3789,18 +3789,18 @@ panfrost_mtk_detile_compute(struct panfrost_context *ctx, struct pipe_blit_info
 
    struct pipe_constant_buffer saved_const = {};
    struct panfrost_constant_buffer *pbuf =
-      &batch->ctx->constant_buffer[PIPE_SHADER_COMPUTE];
-   void *saved_cso = batch->ctx->uncompiled[PIPE_SHADER_COMPUTE];
+      &batch->ctx->constant_buffer[MESA_SHADER_COMPUTE];
+   void *saved_cso = batch->ctx->uncompiled[MESA_SHADER_COMPUTE];
    void *cso = shader->mtk_tiled.detile_cso;
    util_copy_constant_buffer(&pbuf->cb[0], &saved_const, true);
 
    pipe->bind_compute_state(pipe, cso);
-   pipe->set_constant_buffer(pipe, PIPE_SHADER_COMPUTE, 0, false, &cbuf);
+   pipe->set_constant_buffer(pipe, MESA_SHADER_COMPUTE, 0, false, &cbuf);
 
    panfrost_launch_grid_on_batch(pipe, batch, &grid_info);
 
    pipe->bind_compute_state(pipe, saved_cso);
-   pipe->set_constant_buffer(pipe, PIPE_SHADER_COMPUTE, 0, true, &saved_const);
+   pipe->set_constant_buffer(pipe, MESA_SHADER_COMPUTE, 0, true, &saved_const);
 
    panfrost_resource_restore_format(pan_resource(y_src), &y_src_save);
    panfrost_resource_restore_format(pan_resource(uv_src), &uv_src_save);
diff --git a/src/gallium/drivers/panfrost/pan_context.h b/src/gallium/drivers/panfrost/pan_context.h
index 29e89562411..2fecb5579f1 100644
--- a/src/gallium/drivers/panfrost/pan_context.h
+++ b/src/gallium/drivers/panfrost/pan_context.h
@@ -551,7 +551,7 @@ panfrost_clean_state_3d(struct panfrost_context *ctx)
    ctx->dirty = 0;
 
    for (unsigned i = 0; i < PIPE_SHADER_TYPES; ++i) {
-      if (i != PIPE_SHADER_COMPUTE)
+      if (i != MESA_SHADER_COMPUTE)
          ctx->dirty_shader[i] = 0;
    }
 }
diff --git a/src/gallium/drivers/panfrost/pan_csf.c b/src/gallium/drivers/panfrost/pan_csf.c
index 2b4daeeccbe..ac597ea2824 100644
--- a/src/gallium/drivers/panfrost/pan_csf.c
+++ b/src/gallium/drivers/panfrost/pan_csf.c
@@ -906,7 +906,7 @@ csf_emit_shader_regs(struct panfrost_batch *batch, enum pipe_shader_type stage,
    uint64_t resources = panfrost_emit_resources(batch, stage);
 
    assert(stage == MESA_SHADER_VERTEX || stage == MESA_SHADER_FRAGMENT ||
-          stage == PIPE_SHADER_COMPUTE);
+          stage == MESA_SHADER_COMPUTE);
 
 #if PAN_ARCH >= 12
    unsigned offset = (stage == MESA_SHADER_FRAGMENT) ? 2 : 0;
@@ -928,16 +928,16 @@ GENX(csf_launch_grid)(struct panfrost_batch *batch,
                       const struct pipe_grid_info *info)
 {
    /* Empty compute programs are invalid and don't make sense */
-   if (batch->rsd[PIPE_SHADER_COMPUTE] == 0)
+   if (batch->rsd[MESA_SHADER_COMPUTE] == 0)
       return;
 
    struct panfrost_context *ctx = batch->ctx;
    struct panfrost_device *dev = pan_device(ctx->base.screen);
-   struct panfrost_compiled_shader *cs = ctx->prog[PIPE_SHADER_COMPUTE];
+   struct panfrost_compiled_shader *cs = ctx->prog[MESA_SHADER_COMPUTE];
    struct cs_builder *b = batch->csf.cs.builder;
 
-   csf_emit_shader_regs(batch, PIPE_SHADER_COMPUTE,
-                        batch->rsd[PIPE_SHADER_COMPUTE]);
+   csf_emit_shader_regs(batch, MESA_SHADER_COMPUTE,
+                        batch->rsd[MESA_SHADER_COMPUTE]);
 
    cs_move64_to(b, cs_sr_reg64(b, COMPUTE, TSD_0), batch->tls.gpu);
 
diff --git a/src/gallium/drivers/panfrost/pan_jm.c b/src/gallium/drivers/panfrost/pan_jm.c
index b82724e0833..a71c3096ee1 100644
--- a/src/gallium/drivers/panfrost/pan_jm.c
+++ b/src/gallium/drivers/panfrost/pan_jm.c
@@ -330,14 +330,14 @@ GENX(jm_launch_grid)(struct panfrost_batch *batch,
    }
 
    pan_section_pack(t.cpu, COMPUTE_JOB, DRAW, cfg) {
-      cfg.state = batch->rsd[PIPE_SHADER_COMPUTE];
-      cfg.attributes = batch->attribs[PIPE_SHADER_COMPUTE];
-      cfg.attribute_buffers = batch->attrib_bufs[PIPE_SHADER_COMPUTE];
+      cfg.state = batch->rsd[MESA_SHADER_COMPUTE];
+      cfg.attributes = batch->attribs[MESA_SHADER_COMPUTE];
+      cfg.attribute_buffers = batch->attrib_bufs[MESA_SHADER_COMPUTE];
       cfg.thread_storage = batch->tls.gpu;
-      cfg.uniform_buffers = batch->uniform_buffers[PIPE_SHADER_COMPUTE];
-      cfg.push_uniforms = batch->push_uniforms[PIPE_SHADER_COMPUTE];
-      cfg.textures = batch->textures[PIPE_SHADER_COMPUTE];
-      cfg.samplers = batch->samplers[PIPE_SHADER_COMPUTE];
+      cfg.uniform_buffers = batch->uniform_buffers[MESA_SHADER_COMPUTE];
+      cfg.push_uniforms = batch->push_uniforms[MESA_SHADER_COMPUTE];
+      cfg.textures = batch->textures[MESA_SHADER_COMPUTE];
+      cfg.samplers = batch->samplers[MESA_SHADER_COMPUTE];
    }
 
 #if PAN_ARCH == 4
@@ -346,7 +346,7 @@ GENX(jm_launch_grid)(struct panfrost_batch *batch,
 #endif
 #else
    struct panfrost_context *ctx = batch->ctx;
-   struct panfrost_compiled_shader *cs = ctx->prog[PIPE_SHADER_COMPUTE];
+   struct panfrost_compiled_shader *cs = ctx->prog[MESA_SHADER_COMPUTE];
 
    pan_section_pack(t.cpu, COMPUTE_JOB, PAYLOAD, cfg) {
       cfg.workgroup_size_x = info->block[0];
@@ -357,8 +357,8 @@ GENX(jm_launch_grid)(struct panfrost_batch *batch,
       cfg.workgroup_count_y = num_wg[1];
       cfg.workgroup_count_z = num_wg[2];
 
-      jm_emit_shader_env(batch, &cfg.compute, PIPE_SHADER_COMPUTE,
-                         batch->rsd[PIPE_SHADER_COMPUTE]);
+      jm_emit_shader_env(batch, &cfg.compute, MESA_SHADER_COMPUTE,
+                         batch->rsd[MESA_SHADER_COMPUTE]);
 
       /* Workgroups may be merged if the shader does not use barriers
        * or shared memory. This condition is checked against the
diff --git a/src/gallium/drivers/panfrost/pan_screen.c b/src/gallium/drivers/panfrost/pan_screen.c
index 9160baf6c16..9fe219dfda3 100644
--- a/src/gallium/drivers/panfrost/pan_screen.c
+++ b/src/gallium/drivers/panfrost/pan_screen.c
@@ -489,14 +489,14 @@ panfrost_init_shader_caps(struct panfrost_screen *screen)
    struct panfrost_device *dev = &screen->dev;
    bool is_nofp16 = dev->debug & PAN_DBG_NOFP16;
 
-   for (unsigned i = 0; i <= PIPE_SHADER_COMPUTE; i++) {
+   for (unsigned i = 0; i <= MESA_SHADER_COMPUTE; i++) {
       struct pipe_shader_caps *caps =
          (struct pipe_shader_caps *)&screen->base.shader_caps[i];
 
       switch (i) {
       case MESA_SHADER_VERTEX:
       case MESA_SHADER_FRAGMENT:
-      case PIPE_SHADER_COMPUTE:
+      case MESA_SHADER_COMPUTE:
          break;
       default:
          continue;
diff --git a/src/gallium/drivers/panfrost/pan_shader.c b/src/gallium/drivers/panfrost/pan_shader.c
index 0cc68e360e7..f675d5b4f5c 100644
--- a/src/gallium/drivers/panfrost/pan_shader.c
+++ b/src/gallium/drivers/panfrost/pan_shader.c
@@ -406,7 +406,7 @@ panfrost_update_shader_variant(struct panfrost_context *ctx,
                                enum pipe_shader_type type)
 {
    /* No shader variants for compute */
-   if (type == PIPE_SHADER_COMPUTE)
+   if (type == MESA_SHADER_COMPUTE)
       return;
 
    /* We need linking information, defer this */
@@ -610,9 +610,9 @@ panfrost_bind_compute_state(struct pipe_context *pipe, void *cso)
    struct panfrost_context *ctx = pan_context(pipe);
    struct panfrost_uncompiled_shader *uncompiled = cso;
 
-   ctx->uncompiled[PIPE_SHADER_COMPUTE] = uncompiled;
+   ctx->uncompiled[MESA_SHADER_COMPUTE] = uncompiled;
 
-   ctx->prog[PIPE_SHADER_COMPUTE] =
+   ctx->prog[MESA_SHADER_COMPUTE] =
       uncompiled ? util_dynarray_begin(&uncompiled->variants) : NULL;
 }
 
diff --git a/src/gallium/drivers/r600/evergreen_compute.c b/src/gallium/drivers/r600/evergreen_compute.c
index fa36b4c979b..de1f9121d8c 100644
--- a/src/gallium/drivers/r600/evergreen_compute.c
+++ b/src/gallium/drivers/r600/evergreen_compute.c
@@ -146,7 +146,7 @@ static void evergreen_cs_set_constant_buffer(struct r600_context *rctx,
 	cb.buffer = buffer;
 	cb.user_buffer = NULL;
 
-	rctx->b.b.set_constant_buffer(&rctx->b.b, PIPE_SHADER_COMPUTE, cb_index, false, &cb);
+	rctx->b.b.set_constant_buffer(&rctx->b.b, MESA_SHADER_COMPUTE, cb_index, false, &cb);
 }
 
 /* We need to define these R600 registers here, because we can't include
@@ -169,7 +169,7 @@ static void *evergreen_create_compute_state(struct pipe_context *ctx,
 
 	shader->ctx = rctx;
 	shader->local_size = cso->static_shared_mem;
-	shader->sel = r600_create_shader_state_tokens(ctx, cso->prog, cso->ir_type, PIPE_SHADER_COMPUTE);
+	shader->sel = r600_create_shader_state_tokens(ctx, cso->prog, cso->ir_type, MESA_SHADER_COMPUTE);
 
 	/* Precompile the shader with the expected shader key, to reduce jank at
 	 * draw time. Also produces output for shader-db.
@@ -336,7 +336,7 @@ static void compute_emit_cs(struct r600_context *rctx,
 		rctx->cs_block_grid_sizes[i + 4] = info->indirect ? indirect_grid[i] : info->grid[i];
 	}
 	rctx->cs_block_grid_sizes[3] = rctx->cs_block_grid_sizes[7] = 0;
-	rctx->driver_consts[PIPE_SHADER_COMPUTE].cs_block_grid_size_dirty = true;
+	rctx->driver_consts[MESA_SHADER_COMPUTE].cs_block_grid_size_dirty = true;
 
 	if (rctx->b.gfx_level == CAYMAN)
 		global_atomic_count = cayman_emit_atomic_buffer_setup_count(rctx, current, combined_atomics, global_atomic_count);
@@ -346,7 +346,7 @@ static void compute_emit_cs(struct r600_context *rctx,
 	r600_need_cs_space(rctx, 0, true, global_atomic_count);
 
 	if (need_buf_const) {
-		eg_setup_buffer_constants(rctx, PIPE_SHADER_COMPUTE);
+		eg_setup_buffer_constants(rctx, MESA_SHADER_COMPUTE);
 	}
 	r600_update_driver_const_buffers(rctx, true);
 
@@ -384,13 +384,13 @@ static void compute_emit_cs(struct r600_context *rctx,
 	r600_emit_atom(rctx, &rctx->b.render_cond_atom);
 
 	/* Emit constant buffer state */
-	r600_emit_atom(rctx, &rctx->constbuf_state[PIPE_SHADER_COMPUTE].atom);
+	r600_emit_atom(rctx, &rctx->constbuf_state[MESA_SHADER_COMPUTE].atom);
 
 	/* Emit sampler state */
-	r600_emit_atom(rctx, &rctx->samplers[PIPE_SHADER_COMPUTE].states.atom);
+	r600_emit_atom(rctx, &rctx->samplers[MESA_SHADER_COMPUTE].states.atom);
 
 	/* Emit sampler view (texture resource) state */
-	r600_emit_atom(rctx, &rctx->samplers[PIPE_SHADER_COMPUTE].views.atom);
+	r600_emit_atom(rctx, &rctx->samplers[MESA_SHADER_COMPUTE].views.atom);
 
 	/* Emit images state */
 	r600_emit_atom(rctx, &rctx->compute_images.atom);
diff --git a/src/gallium/drivers/r600/evergreen_state.c b/src/gallium/drivers/r600/evergreen_state.c
index 164a599773f..ac215f21052 100644
--- a/src/gallium/drivers/r600/evergreen_state.c
+++ b/src/gallium/drivers/r600/evergreen_state.c
@@ -2387,7 +2387,7 @@ static void evergreen_emit_ps_constant_buffers(struct r600_context *rctx, struct
 
 static void evergreen_emit_cs_constant_buffers(struct r600_context *rctx, struct r600_atom *atom)
 {
-	evergreen_emit_constant_buffers(rctx, &rctx->constbuf_state[PIPE_SHADER_COMPUTE],
+	evergreen_emit_constant_buffers(rctx, &rctx->constbuf_state[MESA_SHADER_COMPUTE],
 					EG_FETCH_CONSTANTS_OFFSET_CS,
 					R_028FC0_ALU_CONST_BUFFER_SIZE_LS_0,
 					R_028F40_ALU_CONST_CACHE_LS_0,
@@ -2513,7 +2513,7 @@ static void evergreen_emit_ps_sampler_views(struct r600_context *rctx, struct r6
 
 static void evergreen_emit_cs_sampler_views(struct r600_context *rctx, struct r600_atom *atom)
 {
-	evergreen_emit_sampler_views(rctx, &rctx->samplers[PIPE_SHADER_COMPUTE].views,
+	evergreen_emit_sampler_views(rctx, &rctx->samplers[MESA_SHADER_COMPUTE].views,
 	                             EG_FETCH_CONSTANTS_OFFSET_CS + R600_MAX_CONST_BUFFERS, RADEON_CP_PACKET3_COMPUTE_MODE);
 }
 
@@ -2882,7 +2882,7 @@ static void evergreen_emit_ps_sampler_states(struct r600_context *rctx, struct r
 
 static void evergreen_emit_cs_sampler_states(struct r600_context *rctx, struct r600_atom *atom)
 {
-	evergreen_emit_sampler_states(rctx, &rctx->samplers[PIPE_SHADER_COMPUTE], 90,
+	evergreen_emit_sampler_states(rctx, &rctx->samplers[MESA_SHADER_COMPUTE], 90,
 	                              R_00A464_TD_CS_SAMPLER0_BORDER_INDEX,
 	                              RADEON_CP_PACKET3_COMPUTE_MODE);
 }
@@ -4438,12 +4438,12 @@ static void evergreen_set_shader_buffers(struct pipe_context *ctx,
 	unsigned old_mask;
 
 	if ((shader != MESA_SHADER_FRAGMENT &&
-        shader != PIPE_SHADER_COMPUTE) || count == 0)
+        shader != MESA_SHADER_COMPUTE) || count == 0)
 		return;
 
 	if (shader == MESA_SHADER_FRAGMENT)
 		istate = &rctx->fragment_buffers;
-	else if (shader == PIPE_SHADER_COMPUTE)
+	else if (shader == MESA_SHADER_COMPUTE)
 		istate = &rctx->compute_buffers;
 
 	old_mask = istate->enabled_mask;
@@ -4539,17 +4539,17 @@ static void evergreen_set_shader_images(struct pipe_context *ctx,
 	unsigned old_mask;
 	struct r600_image_state *istate = NULL;
 	int idx;
-	if (shader != MESA_SHADER_FRAGMENT && shader != PIPE_SHADER_COMPUTE)
+	if (shader != MESA_SHADER_FRAGMENT && shader != MESA_SHADER_COMPUTE)
 		return;
 	if (!count && !unbind_num_trailing_slots)
 		return;
 
 	if (shader == MESA_SHADER_FRAGMENT)
 		istate = &rctx->fragment_images;
-	else if (shader == PIPE_SHADER_COMPUTE)
+	else if (shader == MESA_SHADER_COMPUTE)
 		istate = &rctx->compute_images;
 
-	assert (shader == MESA_SHADER_FRAGMENT || shader == PIPE_SHADER_COMPUTE);
+	assert (shader == MESA_SHADER_FRAGMENT || shader == MESA_SHADER_COMPUTE);
 
 	old_mask = istate->enabled_mask;
 	for (i = start_slot, idx = 0; i < start_slot + count; i++, idx++) {
@@ -4756,7 +4756,7 @@ static void evergreen_get_shader_buffers(struct r600_context *rctx,
 					 uint start_slot, uint count,
 					 struct pipe_shader_buffer *sbuf)
 {
-	assert(shader == PIPE_SHADER_COMPUTE);
+	assert(shader == MESA_SHADER_COMPUTE);
 	int idx, i;
 	struct r600_image_state *istate = &rctx->compute_buffers;
 	struct r600_image_view *rview;
@@ -4789,9 +4789,9 @@ static void evergreen_save_qbo_state(struct pipe_context *ctx, struct r600_qbo_s
 	st->saved_compute = rctx->cs_shader_state.shader;
 
 	/* save constant buffer 0 */
-	evergreen_get_pipe_constant_buffer(rctx, PIPE_SHADER_COMPUTE, 0, &st->saved_const0);
+	evergreen_get_pipe_constant_buffer(rctx, MESA_SHADER_COMPUTE, 0, &st->saved_const0);
 	/* save ssbo 0 */
-	evergreen_get_shader_buffers(rctx, PIPE_SHADER_COMPUTE, 0, 3, st->saved_ssbo);
+	evergreen_get_shader_buffers(rctx, MESA_SHADER_COMPUTE, 0, 3, st->saved_ssbo);
 }
 
 
@@ -4823,7 +4823,7 @@ void evergreen_init_state_functions(struct r600_context *rctx)
 	r600_init_atom(rctx, &rctx->constbuf_state[MESA_SHADER_FRAGMENT].atom, id++, evergreen_emit_ps_constant_buffers, 0);
 	r600_init_atom(rctx, &rctx->constbuf_state[MESA_SHADER_TESS_CTRL].atom, id++, evergreen_emit_tcs_constant_buffers, 0);
 	r600_init_atom(rctx, &rctx->constbuf_state[MESA_SHADER_TESS_EVAL].atom, id++, evergreen_emit_tes_constant_buffers, 0);
-	r600_init_atom(rctx, &rctx->constbuf_state[PIPE_SHADER_COMPUTE].atom, id++, evergreen_emit_cs_constant_buffers, 0);
+	r600_init_atom(rctx, &rctx->constbuf_state[MESA_SHADER_COMPUTE].atom, id++, evergreen_emit_cs_constant_buffers, 0);
 	/* shader program */
 	r600_init_atom(rctx, &rctx->cs_shader_state.atom, id++, evergreen_emit_cs_shader, 0);
 	/* sampler */
@@ -4832,7 +4832,7 @@ void evergreen_init_state_functions(struct r600_context *rctx)
 	r600_init_atom(rctx, &rctx->samplers[MESA_SHADER_TESS_CTRL].states.atom, id++, evergreen_emit_tcs_sampler_states, 0);
 	r600_init_atom(rctx, &rctx->samplers[MESA_SHADER_TESS_EVAL].states.atom, id++, evergreen_emit_tes_sampler_states, 0);
 	r600_init_atom(rctx, &rctx->samplers[MESA_SHADER_FRAGMENT].states.atom, id++, evergreen_emit_ps_sampler_states, 0);
-	r600_init_atom(rctx, &rctx->samplers[PIPE_SHADER_COMPUTE].states.atom, id++, evergreen_emit_cs_sampler_states, 0);
+	r600_init_atom(rctx, &rctx->samplers[MESA_SHADER_COMPUTE].states.atom, id++, evergreen_emit_cs_sampler_states, 0);
 	/* resources */
 	r600_init_atom(rctx, &rctx->vertex_buffer_state.atom, id++, evergreen_fs_emit_vertex_buffers, 0);
 	r600_init_atom(rctx, &rctx->cs_vertex_buffer_state.atom, id++, evergreen_cs_emit_vertex_buffers, 0);
@@ -4841,7 +4841,7 @@ void evergreen_init_state_functions(struct r600_context *rctx)
 	r600_init_atom(rctx, &rctx->samplers[MESA_SHADER_TESS_CTRL].views.atom, id++, evergreen_emit_tcs_sampler_views, 0);
 	r600_init_atom(rctx, &rctx->samplers[MESA_SHADER_TESS_EVAL].views.atom, id++, evergreen_emit_tes_sampler_views, 0);
 	r600_init_atom(rctx, &rctx->samplers[MESA_SHADER_FRAGMENT].views.atom, id++, evergreen_emit_ps_sampler_views, 0);
-	r600_init_atom(rctx, &rctx->samplers[PIPE_SHADER_COMPUTE].views.atom, id++, evergreen_emit_cs_sampler_views, 0);
+	r600_init_atom(rctx, &rctx->samplers[MESA_SHADER_COMPUTE].views.atom, id++, evergreen_emit_cs_sampler_views, 0);
 
 	r600_init_atom(rctx, &rctx->vgt_state.atom, id++, r600_emit_vgt_state, 10);
 
diff --git a/src/gallium/drivers/r600/r600_pipe.c b/src/gallium/drivers/r600/r600_pipe.c
index 3e5ece81b62..a9e25ef9b12 100644
--- a/src/gallium/drivers/r600/r600_pipe.c
+++ b/src/gallium/drivers/r600/r600_pipe.c
@@ -273,14 +273,14 @@ fail:
 
 static void r600_init_shader_caps(struct r600_screen *rscreen)
 {
-	for (unsigned i = 0; i <= PIPE_SHADER_COMPUTE; i++) {
+	for (unsigned i = 0; i <= MESA_SHADER_COMPUTE; i++) {
 		struct pipe_shader_caps *caps =
 			(struct pipe_shader_caps *)&rscreen->b.b.shader_caps[i];
 
 		switch (i) {
 		case MESA_SHADER_TESS_CTRL:
 		case MESA_SHADER_TESS_EVAL:
-		case PIPE_SHADER_COMPUTE:
+		case MESA_SHADER_COMPUTE:
 			if (rscreen->b.family < CHIP_CEDAR)
 				continue;
 			break;
@@ -297,7 +297,7 @@ static void r600_init_shader_caps(struct r600_screen *rscreen)
 		caps->max_outputs = i == MESA_SHADER_FRAGMENT ? 8 : 32;
 		caps->max_temps = 256; /* Max native temporaries. */
 
-		caps->max_const_buffer0_size = i == PIPE_SHADER_COMPUTE ?
+		caps->max_const_buffer0_size = i == MESA_SHADER_COMPUTE ?
 			MIN2(rscreen->b.b.compute_caps.max_mem_alloc_size, INT_MAX) :
 			R600_MAX_CONST_BUFFER_SIZE;
 
@@ -316,7 +316,7 @@ static void r600_init_shader_caps(struct r600_screen *rscreen)
 		caps->max_shader_buffers =
 		caps->max_shader_images =
 			rscreen->b.family >= CHIP_CEDAR &&
-			(i == MESA_SHADER_FRAGMENT || i == PIPE_SHADER_COMPUTE) ? 8 : 0;
+			(i == MESA_SHADER_FRAGMENT || i == MESA_SHADER_COMPUTE) ? 8 : 0;
 
 		if (rscreen->b.family >= CHIP_CEDAR &&
 		    rscreen->has_atomics) {
diff --git a/src/gallium/drivers/r600/r600_pipe_common.h b/src/gallium/drivers/r600/r600_pipe_common.h
index de6ee8baa9c..9774ef3575c 100644
--- a/src/gallium/drivers/r600/r600_pipe_common.h
+++ b/src/gallium/drivers/r600/r600_pipe_common.h
@@ -53,7 +53,7 @@ struct u_log_context;
 #define DBG_GS			(1 << MESA_SHADER_GEOMETRY)
 #define DBG_TCS			(1 << MESA_SHADER_TESS_CTRL)
 #define DBG_TES			(1 << MESA_SHADER_TESS_EVAL)
-#define DBG_CS			(1 << PIPE_SHADER_COMPUTE)
+#define DBG_CS			(1 << MESA_SHADER_COMPUTE)
 #define DBG_ALL_SHADERS		(DBG_FS - 1)
 #define DBG_FS			(1 << 6) /* fetch shader */
 #define DBG_TEX			(1 << 7)
diff --git a/src/gallium/drivers/r600/r600_query.c b/src/gallium/drivers/r600/r600_query.c
index 9d40578066f..079e95cb8dd 100644
--- a/src/gallium/drivers/r600/r600_query.c
+++ b/src/gallium/drivers/r600/r600_query.c
@@ -1579,8 +1579,8 @@ static void r600_restore_qbo_state(struct r600_common_context *rctx,
 				   struct r600_qbo_state *st)
 {
 	rctx->b.bind_compute_state(&rctx->b, st->saved_compute);
-	rctx->b.set_constant_buffer(&rctx->b, PIPE_SHADER_COMPUTE, 0, true, &st->saved_const0);
-	rctx->b.set_shader_buffers(&rctx->b, PIPE_SHADER_COMPUTE, 0, 3, st->saved_ssbo, ~0);
+	rctx->b.set_constant_buffer(&rctx->b, MESA_SHADER_COMPUTE, 0, true, &st->saved_const0);
+	rctx->b.set_shader_buffers(&rctx->b, MESA_SHADER_COMPUTE, 0, 3, st->saved_ssbo, ~0);
 	for (unsigned i = 0; i < 3; ++i)
 		pipe_resource_reference(&st->saved_ssbo[i].buffer, NULL);
 }
@@ -1712,9 +1712,9 @@ static void r600_query_hw_get_result_resource(struct r600_common_context *rctx,
 		} else
 			consts.buffer_offset = 0;
 
-		rctx->b.set_constant_buffer(&rctx->b, PIPE_SHADER_COMPUTE, 0, false, &constant_buffer);
+		rctx->b.set_constant_buffer(&rctx->b, MESA_SHADER_COMPUTE, 0, false, &constant_buffer);
 
-		rctx->b.set_shader_buffers(&rctx->b, PIPE_SHADER_COMPUTE, 0, 3, ssbo, ~0);
+		rctx->b.set_shader_buffers(&rctx->b, MESA_SHADER_COMPUTE, 0, 3, ssbo, ~0);
 
 		if ((flags & PIPE_QUERY_WAIT) && qbuf == &query->buffer) {
 			uint64_t va;
diff --git a/src/gallium/drivers/r600/r600_shader.c b/src/gallium/drivers/r600/r600_shader.c
index 4d8a4a1b9d8..738cb829fe1 100644
--- a/src/gallium/drivers/r600/r600_shader.c
+++ b/src/gallium/drivers/r600/r600_shader.c
@@ -268,7 +268,7 @@ int r600_pipe_shader_create(struct pipe_context *ctx,
 			r600_update_ps_state(ctx, shader);
 		}
 		break;
-	case PIPE_SHADER_COMPUTE:
+	case MESA_SHADER_COMPUTE:
 		evergreen_update_ls_state(ctx, shader);
 		break;
 	default:
diff --git a/src/gallium/drivers/r600/r600_state_common.c b/src/gallium/drivers/r600/r600_state_common.c
index 5212b905c41..19a93e42454 100644
--- a/src/gallium/drivers/r600/r600_state_common.c
+++ b/src/gallium/drivers/r600/r600_state_common.c
@@ -746,7 +746,7 @@ static int r600_get_hw_atomic_count(const struct pipe_context *ctx,
 	int value = 0;
 	switch (shader) {
 	case MESA_SHADER_FRAGMENT:
-	case PIPE_SHADER_COMPUTE:
+	case MESA_SHADER_COMPUTE:
 	default:
 		break;
 	case MESA_SHADER_VERTEX:
@@ -840,7 +840,7 @@ static inline void r600_shader_selector_key(const struct pipe_context *ctx,
 		key->tcs.prim_mode = rctx->tes_shader->info.properties[TGSI_PROPERTY_TES_PRIM_MODE];
 		key->tcs.first_atomic_counter = r600_get_hw_atomic_count(ctx, MESA_SHADER_TESS_CTRL);
 		break;
-	case PIPE_SHADER_COMPUTE:
+	case MESA_SHADER_COMPUTE:
 		break;
 	default:
 		assert(0);
@@ -881,7 +881,7 @@ r600_shader_precompile_key(const struct pipe_context *ctx,
 		key->tcs.prim_mode = MESA_PRIM_TRIANGLES;
 		break;
 
-	case PIPE_SHADER_COMPUTE:
+	case MESA_SHADER_COMPUTE:
 		break;
 
 	default:
@@ -1323,8 +1323,8 @@ void r600_update_driver_const_buffers(struct r600_context *rctx, bool compute_on
 	struct pipe_constant_buffer cb;
 	int start, end;
 
-	start = compute_only ? PIPE_SHADER_COMPUTE : 0;
-	end = compute_only ? PIPE_SHADER_TYPES : PIPE_SHADER_COMPUTE;
+	start = compute_only ? MESA_SHADER_COMPUTE : 0;
+	end = compute_only ? PIPE_SHADER_TYPES : MESA_SHADER_COMPUTE;
 
 	int last_vertex_stage = MESA_SHADER_VERTEX;
 	if (rctx->tes_shader)
@@ -1368,7 +1368,7 @@ void r600_update_driver_const_buffers(struct r600_context *rctx, bool compute_on
 		}
 
 		else if (info->cs_block_grid_size_dirty) {
-			assert(sh == PIPE_SHADER_COMPUTE);
+			assert(sh == MESA_SHADER_COMPUTE);
 			if (!size) {
 				ptr = rctx->cs_block_grid_sizes;
 				size = R600_CS_BLOCK_GRID_SIZE;
@@ -1399,7 +1399,7 @@ void r600_update_driver_const_buffers(struct r600_context *rctx, bool compute_on
 				memcpy(ptr, rctx->clip_state.state.ucp, R600_UCP_SIZE);
 			if (sh == MESA_SHADER_FRAGMENT)
 				memcpy(ptr, rctx->sample_positions, R600_UCP_SIZE);
-			if (sh == PIPE_SHADER_COMPUTE)
+			if (sh == MESA_SHADER_COMPUTE)
 				memcpy(ptr, rctx->cs_block_grid_sizes, R600_CS_BLOCK_GRID_SIZE);
 			if (sh == MESA_SHADER_TESS_CTRL)
 				memcpy(ptr, rctx->tess_state, R600_TCS_DEFAULT_LEVELS_SIZE);
@@ -1497,7 +1497,7 @@ void eg_setup_buffer_constants(struct r600_context *rctx, int shader_type)
 
 	if (shader_type == MESA_SHADER_FRAGMENT) {
 		images = &rctx->fragment_images;
-	} else if (shader_type == PIPE_SHADER_COMPUTE) {
+	} else if (shader_type == MESA_SHADER_COMPUTE) {
 		images = &rctx->compute_images;
 	}
 
@@ -1673,7 +1673,7 @@ void r600_update_compressed_resource_state(struct r600_context *rctx, bool compu
 		rctx->b.last_compressed_colortex_counter = counter;
 
 		if (compute_only) {
-			r600_update_compressed_colortex_mask(&rctx->samplers[PIPE_SHADER_COMPUTE].views);
+			r600_update_compressed_colortex_mask(&rctx->samplers[MESA_SHADER_COMPUTE].views);
 		} else {
 			for (i = 0; i < PIPE_SHADER_TYPES; ++i) {
 				r600_update_compressed_colortex_mask(&rctx->samplers[i].views);
@@ -1689,7 +1689,7 @@ void r600_update_compressed_resource_state(struct r600_context *rctx, bool compu
 		struct r600_samplerview_state *views = &rctx->samplers[i].views;
 
 		if (compute_only)
-			if (i != PIPE_SHADER_COMPUTE)
+			if (i != MESA_SHADER_COMPUTE)
 				continue;
 		if (views->compressed_depthtex_mask) {
 			r600_decompress_depth_textures(rctx, views);
diff --git a/src/gallium/drivers/r600/sfn/sfn_shader_cs.cpp b/src/gallium/drivers/r600/sfn/sfn_shader_cs.cpp
index 5bd25d2dd08..ac10b3b2324 100644
--- a/src/gallium/drivers/r600/sfn/sfn_shader_cs.cpp
+++ b/src/gallium/drivers/r600/sfn/sfn_shader_cs.cpp
@@ -59,7 +59,7 @@ ComputeShader::process_stage_intrinsic(nir_intrinsic_instr *instr)
 void
 ComputeShader::do_get_shader_info(r600_shader *sh_info)
 {
-   sh_info->processor_type = PIPE_SHADER_COMPUTE;
+   sh_info->processor_type = MESA_SHADER_COMPUTE;
 }
 
 bool
diff --git a/src/gallium/drivers/radeonsi/gfx11_query.c b/src/gallium/drivers/radeonsi/gfx11_query.c
index 20558ba875e..790da31bb87 100644
--- a/src/gallium/drivers/radeonsi/gfx11_query.c
+++ b/src/gallium/drivers/radeonsi/gfx11_query.c
@@ -374,7 +374,7 @@ static void gfx11_sh_query_get_result_resource(struct si_context *sctx, struct s
          ssbo[2].buffer_size = is_result_64bit ? 8 : 4;
       }
 
-      sctx->b.set_constant_buffer(&sctx->b, PIPE_SHADER_COMPUTE, 0, false, &constant_buffer);
+      sctx->b.set_constant_buffer(&sctx->b, MESA_SHADER_COMPUTE, 0, false, &constant_buffer);
 
       if (flags & PIPE_QUERY_WAIT) {
          uint64_t va;
diff --git a/src/gallium/drivers/radeonsi/si_barrier.c b/src/gallium/drivers/radeonsi/si_barrier.c
index dcccb559301..b31a3d444ad 100644
--- a/src/gallium/drivers/radeonsi/si_barrier.c
+++ b/src/gallium/drivers/radeonsi/si_barrier.c
@@ -513,10 +513,10 @@ void si_barrier_before_internal_op(struct si_context *sctx, unsigned flags,
                             SI_BIND_SHADER_BUFFER(MESA_SHADER_FRAGMENT) |
                             SI_BIND_IMAGE_BUFFER(MESA_SHADER_FRAGMENT) |
                             SI_BIND_SAMPLER_BUFFER(MESA_SHADER_FRAGMENT);
-   const unsigned cs_mask = SI_BIND_CONSTANT_BUFFER(PIPE_SHADER_COMPUTE) |
-                            SI_BIND_SHADER_BUFFER(PIPE_SHADER_COMPUTE) |
-                            SI_BIND_IMAGE_BUFFER(PIPE_SHADER_COMPUTE) |
-                            SI_BIND_SAMPLER_BUFFER(PIPE_SHADER_COMPUTE);
+   const unsigned cs_mask = SI_BIND_CONSTANT_BUFFER(MESA_SHADER_COMPUTE) |
+                            SI_BIND_SHADER_BUFFER(MESA_SHADER_COMPUTE) |
+                            SI_BIND_IMAGE_BUFFER(MESA_SHADER_COMPUTE) |
+                            SI_BIND_SAMPLER_BUFFER(MESA_SHADER_COMPUTE);
 
    for (unsigned i = 0; i < num_buffers; i++) {
       struct si_resource *buf = si_resource(buffers[i].buffer);
diff --git a/src/gallium/drivers/radeonsi/si_compute.c b/src/gallium/drivers/radeonsi/si_compute.c
index 61942a926a9..ecdc35ee00f 100644
--- a/src/gallium/drivers/radeonsi/si_compute.c
+++ b/src/gallium/drivers/radeonsi/si_compute.c
@@ -164,9 +164,9 @@ static void *si_create_compute_state(struct pipe_context *ctx, const struct pipe
    sel->screen = sscreen;
    simple_mtx_init(&sel->mutex, mtx_plain);
    sel->const_and_shader_buf_descriptors_index =
-      si_const_and_shader_buffer_descriptors_idx(PIPE_SHADER_COMPUTE);
+      si_const_and_shader_buffer_descriptors_idx(MESA_SHADER_COMPUTE);
    sel->sampler_and_images_descriptors_index =
-      si_sampler_and_image_descriptors_idx(PIPE_SHADER_COMPUTE);
+      si_sampler_and_image_descriptors_idx(MESA_SHADER_COMPUTE);
    sel->info.base.shared_size = cso->static_shared_mem;
    program->shader.selector = &program->sel;
 
@@ -220,7 +220,7 @@ static void si_bind_compute_state(struct pipe_context *ctx, void *state)
    /* Wait because we need active slot usage masks. */
    util_queue_fence_wait(&sel->ready);
 
-   si_update_common_shader_state(sctx, sel, PIPE_SHADER_COMPUTE);
+   si_update_common_shader_state(sctx, sel, MESA_SHADER_COMPUTE);
 
    sctx->compute_shaderbuf_sgprs_dirty = true;
    sctx->compute_image_sgprs_dirty = true;
@@ -842,7 +842,7 @@ static bool si_check_needs_implicit_sync(struct si_context *sctx, uint32_t usage
     * TODO: Bindless textures are not handled, and thus are not synchronized.
     */
    struct si_shader_info *info = &sctx->cs_shader_state.program->sel.info;
-   struct si_samplers *samplers = &sctx->samplers[PIPE_SHADER_COMPUTE];
+   struct si_samplers *samplers = &sctx->samplers[MESA_SHADER_COMPUTE];
    unsigned mask = samplers->enabled_mask & info->base.textures_used;
 
    while (mask) {
@@ -854,7 +854,7 @@ static bool si_check_needs_implicit_sync(struct si_context *sctx, uint32_t usage
          return true;
    }
 
-   struct si_images *images = &sctx->images[PIPE_SHADER_COMPUTE];
+   struct si_images *images = &sctx->images[MESA_SHADER_COMPUTE];
    mask = BITFIELD_MASK(info->base.num_images) & images->enabled_mask;
 
    while (mask) {
@@ -907,9 +907,9 @@ static void si_launch_grid(struct pipe_context *ctx, const struct pipe_grid_info
       }
 
       if (sctx->gfx_level < GFX11)
-         gfx6_decompress_textures(sctx, 1 << PIPE_SHADER_COMPUTE);
+         gfx6_decompress_textures(sctx, 1 << MESA_SHADER_COMPUTE);
       else if (sctx->gfx_level < GFX12)
-         gfx11_decompress_textures(sctx, 1 << PIPE_SHADER_COMPUTE);
+         gfx11_decompress_textures(sctx, 1 << MESA_SHADER_COMPUTE);
    }
 
    if (info->indirect) {
@@ -996,11 +996,11 @@ static void si_launch_grid(struct pipe_context *ctx, const struct pipe_grid_info
 
    if (sctx->gfx_level < GFX12) {
       /* Mark displayable DCC as dirty for bound images. */
-      unsigned display_dcc_store_mask = sctx->images[PIPE_SHADER_COMPUTE].display_dcc_store_mask &
+      unsigned display_dcc_store_mask = sctx->images[MESA_SHADER_COMPUTE].display_dcc_store_mask &
                                   BITFIELD_MASK(program->sel.info.base.num_images);
       while (display_dcc_store_mask) {
          struct si_texture *tex = (struct si_texture *)
-            sctx->images[PIPE_SHADER_COMPUTE].views[u_bit_scan(&display_dcc_store_mask)].resource;
+            sctx->images[MESA_SHADER_COMPUTE].views[u_bit_scan(&display_dcc_store_mask)].resource;
 
          si_mark_display_dcc_dirty(sctx, tex);
       }
diff --git a/src/gallium/drivers/radeonsi/si_compute_blit.c b/src/gallium/drivers/radeonsi/si_compute_blit.c
index e17c579fa1e..486d9baf922 100644
--- a/src/gallium/drivers/radeonsi/si_compute_blit.c
+++ b/src/gallium/drivers/radeonsi/si_compute_blit.c
@@ -62,17 +62,17 @@ void si_launch_grid_internal_ssbos(struct si_context *sctx, struct pipe_grid_inf
    /* Save states. */
    struct pipe_shader_buffer saved_sb[3] = {};
    assert(num_buffers <= ARRAY_SIZE(saved_sb));
-   si_get_shader_buffers(sctx, PIPE_SHADER_COMPUTE, 0, num_buffers, saved_sb);
+   si_get_shader_buffers(sctx, MESA_SHADER_COMPUTE, 0, num_buffers, saved_sb);
 
    unsigned saved_writable_mask = 0;
    for (unsigned i = 0; i < num_buffers; i++) {
-      if (sctx->const_and_shader_buffers[PIPE_SHADER_COMPUTE].writable_mask &
+      if (sctx->const_and_shader_buffers[MESA_SHADER_COMPUTE].writable_mask &
           (1u << si_get_shaderbuf_slot(i)))
          saved_writable_mask |= 1 << i;
    }
 
    /* Bind buffers and launch compute. */
-   si_set_shader_buffers(&sctx->b, PIPE_SHADER_COMPUTE, 0, num_buffers, buffers,
+   si_set_shader_buffers(&sctx->b, MESA_SHADER_COMPUTE, 0, num_buffers, buffers,
                          writeable_bitmask,
                          true /* don't update bind_history to prevent unnecessary syncs later */);
 
@@ -81,7 +81,7 @@ void si_launch_grid_internal_ssbos(struct si_context *sctx, struct pipe_grid_inf
    si_compute_end_internal(sctx);
 
    /* Restore states. */
-   sctx->b.set_shader_buffers(&sctx->b, PIPE_SHADER_COMPUTE, 0, num_buffers, saved_sb,
+   sctx->b.set_shader_buffers(&sctx->b, MESA_SHADER_COMPUTE, 0, num_buffers, saved_sb,
                               saved_writable_mask);
    for (int i = 0; i < num_buffers; i++)
       pipe_resource_reference(&saved_sb[i].buffer, NULL);
@@ -323,19 +323,19 @@ static void si_compute_save_and_bind_images(struct si_context *sctx, unsigned nu
       }
 
       /* Save the image. */
-      util_copy_image_view(&saved_images[i], &sctx->images[PIPE_SHADER_COMPUTE].views[i]);
+      util_copy_image_view(&saved_images[i], &sctx->images[MESA_SHADER_COMPUTE].views[i]);
    }
 
    /* This must be before the barrier and si_compute_begin_internal because it might invoke DCC
     * decompression.
     */
-   sctx->b.set_shader_images(&sctx->b, PIPE_SHADER_COMPUTE, 0, num_images, 0, images);
+   sctx->b.set_shader_images(&sctx->b, MESA_SHADER_COMPUTE, 0, num_images, 0, images);
 }
 
 static void si_compute_restore_images(struct si_context *sctx, unsigned num_images,
                                       struct pipe_image_view *saved_images)
 {
-   sctx->b.set_shader_images(&sctx->b, PIPE_SHADER_COMPUTE, 0, num_images, 0, saved_images);
+   sctx->b.set_shader_images(&sctx->b, MESA_SHADER_COMPUTE, 0, num_images, 0, saved_images);
    for (unsigned i = 0; i < num_images; i++)
       pipe_resource_reference(&saved_images[i].resource, NULL);
 }
@@ -449,7 +449,7 @@ void si_compute_expand_fmask(struct pipe_context *ctx, struct pipe_resource *tex
 
    /* Save states. */
    struct pipe_image_view saved_image = {0};
-   util_copy_image_view(&saved_image, &sctx->images[PIPE_SHADER_COMPUTE].views[0]);
+   util_copy_image_view(&saved_image, &sctx->images[MESA_SHADER_COMPUTE].views[0]);
 
    /* Bind the image. */
    struct pipe_image_view image = {0};
@@ -461,7 +461,7 @@ void si_compute_expand_fmask(struct pipe_context *ctx, struct pipe_resource *tex
    if (is_array)
       image.u.tex.last_layer = tex->array_size - 1;
 
-   ctx->set_shader_images(ctx, PIPE_SHADER_COMPUTE, 0, 1, 0, &image);
+   ctx->set_shader_images(ctx, MESA_SHADER_COMPUTE, 0, 1, 0, &image);
 
    /* Bind the shader. */
    void **shader = &sctx->cs_fmask_expand[log_samples - 1][is_array];
@@ -479,7 +479,7 @@ void si_compute_expand_fmask(struct pipe_context *ctx, struct pipe_resource *tex
    si_barrier_after_internal_op(sctx, 0, 0, NULL, 0, 1, &image);
 
    /* Restore previous states. */
-   ctx->set_shader_images(ctx, PIPE_SHADER_COMPUTE, 0, 1, 0, &saved_image);
+   ctx->set_shader_images(ctx, MESA_SHADER_COMPUTE, 0, 1, 0, &saved_image);
    pipe_resource_reference(&saved_image.resource, NULL);
 
    /* Array of fully expanded FMASK values, arranged by [log2(fragments)][log2(samples)-1]. */
diff --git a/src/gallium/drivers/radeonsi/si_descriptors.c b/src/gallium/drivers/radeonsi/si_descriptors.c
index 03c434bafab..d230463c59a 100644
--- a/src/gallium/drivers/radeonsi/si_descriptors.c
+++ b/src/gallium/drivers/radeonsi/si_descriptors.c
@@ -520,7 +520,7 @@ static void si_set_sampler_views(struct si_context *sctx, unsigned shader,
    samplers->needs_color_decompress_mask &= ~unbound_mask;
 
    sctx->descriptors_dirty |= 1u << si_sampler_and_image_descriptors_idx(shader);
-   if (shader != PIPE_SHADER_COMPUTE)
+   if (shader != MESA_SHADER_COMPUTE)
       si_mark_atom_dirty(sctx, &sctx->atoms.s.gfx_shader_pointers);
 }
 
@@ -643,7 +643,7 @@ static void si_disable_shader_image(struct si_context *ctx, unsigned shader, uns
       images->enabled_mask &= ~(1u << slot);
       images->display_dcc_store_mask &= ~(1u << slot);
       ctx->descriptors_dirty |= 1u << si_sampler_and_image_descriptors_idx(shader);
-      if (shader != PIPE_SHADER_COMPUTE)
+      if (shader != MESA_SHADER_COMPUTE)
          si_mark_atom_dirty(ctx, &ctx->atoms.s.gfx_shader_pointers);
    }
 }
@@ -785,7 +785,7 @@ static void si_set_shader_image(struct si_context *ctx, unsigned shader, unsigne
             images->display_dcc_store_mask |= 1u << slot;
 
             /* Set displayable_dcc_dirty for non-compute stages conservatively (before draw calls). */
-            if (shader != PIPE_SHADER_COMPUTE)
+            if (shader != MESA_SHADER_COMPUTE)
                tex->displayable_dcc_dirty = true;
          } else {
             images->display_dcc_store_mask &= ~(1u << slot);
@@ -799,7 +799,7 @@ static void si_set_shader_image(struct si_context *ctx, unsigned shader, unsigne
 
    images->enabled_mask |= 1u << slot;
    ctx->descriptors_dirty |= 1u << si_sampler_and_image_descriptors_idx(shader);
-   if (shader != PIPE_SHADER_COMPUTE)
+   if (shader != MESA_SHADER_COMPUTE)
       si_mark_atom_dirty(ctx, &ctx->atoms.s.gfx_shader_pointers);
 
    /* Since this can flush, it must be done after enabled_mask is updated. */
@@ -834,7 +834,7 @@ static void si_set_shader_images(struct pipe_context *pipe, enum pipe_shader_typ
    for (i = 0; i < unbind_num_trailing_slots; ++i, ++slot)
       si_set_shader_image(ctx, shader, slot, NULL, false);
 
-   if (shader == PIPE_SHADER_COMPUTE &&
+   if (shader == MESA_SHADER_COMPUTE &&
        ctx->cs_shader_state.program &&
        start_slot < ctx->cs_shader_state.program->sel.cs_num_images_in_user_sgprs)
       ctx->compute_image_sgprs_dirty = true;
@@ -998,7 +998,7 @@ static void si_bind_sampler_states(struct pipe_context *ctx, enum pipe_shader_ty
       si_set_sampler_state_desc(sstates[i], sview, tex, desc->list + desc_slot * 16 + 12);
 
       sctx->descriptors_dirty |= 1u << si_sampler_and_image_descriptors_idx(shader);
-      if (shader != PIPE_SHADER_COMPUTE)
+      if (shader != MESA_SHADER_COMPUTE)
          si_mark_atom_dirty(sctx, &sctx->atoms.s.gfx_shader_pointers);
    }
 }
@@ -1193,7 +1193,7 @@ void si_get_inline_uniform_state(union si_shader_key *key, enum pipe_shader_type
 
 void si_invalidate_inlinable_uniforms(struct si_context *sctx, enum pipe_shader_type shader)
 {
-   if (shader == PIPE_SHADER_COMPUTE)
+   if (shader == MESA_SHADER_COMPUTE)
       return;
 
    bool inline_uniforms;
@@ -1246,7 +1246,7 @@ static void si_set_inlinable_constants(struct pipe_context *ctx,
 {
    struct si_context *sctx = (struct si_context *)ctx;
 
-   if (shader == PIPE_SHADER_COMPUTE)
+   if (shader == MESA_SHADER_COMPUTE)
       return;
 
    bool inline_uniforms;
@@ -1345,7 +1345,7 @@ void si_set_shader_buffers(struct pipe_context *ctx, enum pipe_shader_type shade
 
    assert(start_slot + count <= SI_NUM_SHADER_BUFFERS);
 
-   if (shader == PIPE_SHADER_COMPUTE &&
+   if (shader == MESA_SHADER_COMPUTE &&
        sctx->cs_shader_state.program &&
        start_slot < sctx->cs_shader_state.program->sel.cs_num_shaderbufs_in_user_sgprs)
       sctx->compute_shaderbuf_sgprs_dirty = true;
@@ -1716,7 +1716,7 @@ void si_rebind_buffer(struct si_context *sctx, struct pipe_resource *buf)
                                        si_const_and_shader_buffer_descriptors_idx(shader),
                                        BITFIELD64_MASK(SI_NUM_SHADER_BUFFERS), buf,
                                        sctx->const_and_shader_buffers[shader].priority) &&
-             shader == PIPE_SHADER_COMPUTE) {
+             shader == MESA_SHADER_COMPUTE) {
             sctx->compute_shaderbuf_sgprs_dirty = true;
          }
       }
@@ -1741,7 +1741,7 @@ void si_rebind_buffer(struct si_context *sctx, struct pipe_resource *buf)
                si_set_buf_desc_address(si_resource(buffer), samplers->views[i]->u.buf.offset,
                                        descs->list + desc_slot * 16);
                sctx->descriptors_dirty |= 1u << si_sampler_and_image_descriptors_idx(shader);
-               if (shader != PIPE_SHADER_COMPUTE)
+               if (shader != MESA_SHADER_COMPUTE)
                   si_mark_atom_dirty(sctx, &sctx->atoms.s.gfx_shader_pointers);
 
                radeon_add_to_buffer_list(sctx, &sctx->gfx_cs, si_resource(buffer), RADEON_USAGE_READ |
@@ -1773,14 +1773,14 @@ void si_rebind_buffer(struct si_context *sctx, struct pipe_resource *buf)
                si_set_buf_desc_address(si_resource(buffer), images->views[i].u.buf.offset,
                                        descs->list + desc_slot * 8);
                sctx->descriptors_dirty |= 1u << si_sampler_and_image_descriptors_idx(shader);
-               if (shader != PIPE_SHADER_COMPUTE)
+               if (shader != MESA_SHADER_COMPUTE)
                   si_mark_atom_dirty(sctx, &sctx->atoms.s.gfx_shader_pointers);
 
                radeon_add_to_buffer_list(sctx, &sctx->gfx_cs, si_resource(buffer),
                                          RADEON_USAGE_READWRITE |
                                          RADEON_PRIO_SAMPLER_BUFFER);
 
-               if (shader == PIPE_SHADER_COMPUTE)
+               if (shader == MESA_SHADER_COMPUTE)
                   sctx->compute_image_sgprs_dirty = true;
             }
          }
@@ -2416,7 +2416,7 @@ void si_emit_compute_shader_pointers(struct si_context *sctx)
    unsigned num_shaderbufs = shader->cs_num_shaderbufs_in_user_sgprs;
 
    if (num_shaderbufs && sctx->compute_shaderbuf_sgprs_dirty) {
-      struct si_descriptors *desc = si_const_and_shader_buffer_descriptors(sctx, PIPE_SHADER_COMPUTE);
+      struct si_descriptors *desc = si_const_and_shader_buffer_descriptors(sctx, MESA_SHADER_COMPUTE);
 
       radeon_set_sh_reg_seq(R_00B900_COMPUTE_USER_DATA_0 +
                             shader->cs_shaderbufs_sgpr_index * 4,
@@ -2431,7 +2431,7 @@ void si_emit_compute_shader_pointers(struct si_context *sctx)
    /* Set image descriptors in user SGPRs. */
    unsigned num_images = shader->cs_num_images_in_user_sgprs;
    if (num_images && sctx->compute_image_sgprs_dirty) {
-      struct si_descriptors *desc = si_sampler_and_image_descriptors(sctx, PIPE_SHADER_COMPUTE);
+      struct si_descriptors *desc = si_sampler_and_image_descriptors(sctx, MESA_SHADER_COMPUTE);
 
       radeon_set_sh_reg_seq(R_00B900_COMPUTE_USER_DATA_0 +
                             shader->cs_images_sgpr_index * 4,
@@ -2853,7 +2853,7 @@ static void si_emit_gfx_resources_add_all_to_bo_list(struct si_context *sctx, un
 void si_init_all_descriptors(struct si_context *sctx)
 {
    int i;
-   unsigned first_shader = sctx->is_gfx_queue ? 0 : PIPE_SHADER_COMPUTE;
+   unsigned first_shader = sctx->is_gfx_queue ? 0 : MESA_SHADER_COMPUTE;
    unsigned hs_sgpr0, gs_sgpr0;
 
    if (sctx->gfx_level >= GFX12) {
@@ -3077,7 +3077,7 @@ static void si_emit_gfx_resources_add_all_to_bo_list(struct si_context *sctx, un
 
 bool si_compute_resources_check_encrypted(struct si_context *sctx)
 {
-   unsigned sh = PIPE_SHADER_COMPUTE;
+   unsigned sh = MESA_SHADER_COMPUTE;
 
    struct si_shader_info* info = &sctx->cs_shader_state.program->sel.info;
 
@@ -3092,7 +3092,7 @@ bool si_compute_resources_check_encrypted(struct si_context *sctx)
 
 void si_compute_resources_add_all_to_bo_list(struct si_context *sctx)
 {
-   unsigned sh = PIPE_SHADER_COMPUTE;
+   unsigned sh = MESA_SHADER_COMPUTE;
 
    si_buffer_resources_begin_new_cs(sctx, &sctx->const_and_shader_buffers[sh]);
    si_sampler_views_begin_new_cs(sctx, &sctx->samplers[sh]);
diff --git a/src/gallium/drivers/radeonsi/si_get.c b/src/gallium/drivers/radeonsi/si_get.c
index 124cfcac3e1..bb1636cfc83 100644
--- a/src/gallium/drivers/radeonsi/si_get.c
+++ b/src/gallium/drivers/radeonsi/si_get.c
@@ -927,7 +927,7 @@ void si_init_screen_get_functions(struct si_screen *sscreen)
 
 void si_init_shader_caps(struct si_screen *sscreen)
 {
-   for (unsigned i = 0; i <= PIPE_SHADER_COMPUTE; i++) {
+   for (unsigned i = 0; i <= MESA_SHADER_COMPUTE; i++) {
       struct pipe_shader_caps *caps =
          (struct pipe_shader_caps *)&sscreen->b.shader_caps[i];
 
diff --git a/src/gallium/drivers/radeonsi/si_pipe.c b/src/gallium/drivers/radeonsi/si_pipe.c
index e00875c7e15..703e7a1b12e 100644
--- a/src/gallium/drivers/radeonsi/si_pipe.c
+++ b/src/gallium/drivers/radeonsi/si_pipe.c
@@ -770,7 +770,7 @@ static struct pipe_context *si_create_context(struct pipe_screen *screen, unsign
       }
       sctx->null_const_buf.buffer_size = sctx->null_const_buf.buffer->width0;
 
-      unsigned start_shader = sctx->is_gfx_queue ? 0 : PIPE_SHADER_COMPUTE;
+      unsigned start_shader = sctx->is_gfx_queue ? 0 : MESA_SHADER_COMPUTE;
       for (shader = start_shader; shader < SI_NUM_SHADERS; shader++) {
          for (i = 0; i < SI_NUM_CONST_BUFFERS; i++) {
             sctx->b.set_constant_buffer(&sctx->b, shader, i, false, &sctx->null_const_buf);
diff --git a/src/gallium/drivers/radeonsi/si_pipe.h b/src/gallium/drivers/radeonsi/si_pipe.h
index 83fbc7aeb8e..09c0f3b4708 100644
--- a/src/gallium/drivers/radeonsi/si_pipe.h
+++ b/src/gallium/drivers/radeonsi/si_pipe.h
@@ -110,7 +110,7 @@ struct ac_llvm_compiler;
 #define SI_RESOURCE_FLAG_32BIT             (PIPE_RESOURCE_FLAG_DRV_PRIV << 6)
 #define SI_RESOURCE_FLAG_CLEAR             (PIPE_RESOURCE_FLAG_DRV_PRIV << 7)
 
-#define SI_SQTT_STATE_DIRTY_BIT            BITFIELD_BIT(PIPE_SHADER_COMPUTE + 1)
+#define SI_SQTT_STATE_DIRTY_BIT            BITFIELD_BIT(MESA_SHADER_COMPUTE + 1)
 
 enum si_has_gs {
    GS_OFF,
diff --git a/src/gallium/drivers/radeonsi/si_query.c b/src/gallium/drivers/radeonsi/si_query.c
index 8a5ac817f3f..d2224c88abe 100644
--- a/src/gallium/drivers/radeonsi/si_query.c
+++ b/src/gallium/drivers/radeonsi/si_query.c
@@ -1619,7 +1619,7 @@ static void si_query_hw_get_result_resource(struct si_context *sctx, struct si_q
          params.start_offset += qbuf->results_end - query->result_size;
       }
 
-      sctx->b.set_constant_buffer(&sctx->b, PIPE_SHADER_COMPUTE, 0, false, &constant_buffer);
+      sctx->b.set_constant_buffer(&sctx->b, MESA_SHADER_COMPUTE, 0, false, &constant_buffer);
 
       ssbo[0].buffer = &qbuf->buf->b.b;
       ssbo[0].buffer_offset = params.start_offset;
diff --git a/src/gallium/drivers/radeonsi/si_sqtt.c b/src/gallium/drivers/radeonsi/si_sqtt.c
index f6c80fb6bae..365d63e226d 100644
--- a/src/gallium/drivers/radeonsi/si_sqtt.c
+++ b/src/gallium/drivers/radeonsi/si_sqtt.c
@@ -722,7 +722,7 @@ si_sqtt_pipe_to_rgp_shader_stage(union si_shader_key *key, enum pipe_shader_type
          return RGP_HW_STAGE_GS;
       case MESA_SHADER_FRAGMENT:
          return RGP_HW_STAGE_PS;
-      case PIPE_SHADER_COMPUTE:
+      case MESA_SHADER_COMPUTE:
          return RGP_HW_STAGE_CS;
       default:
          UNREACHABLE("invalid mesa shader stage");
@@ -752,7 +752,7 @@ si_sqtt_add_code_object(struct si_context *sctx,
       enum rgp_hardware_stages hw_stage;
 
       if (is_compute) {
-         if (i != PIPE_SHADER_COMPUTE)
+         if (i != MESA_SHADER_COMPUTE)
             continue;
          shader = &sctx->cs_shader_state.program->shader;
          hw_stage = RGP_HW_STAGE_CS;
diff --git a/src/gallium/drivers/radeonsi/si_state.c b/src/gallium/drivers/radeonsi/si_state.c
index a3cd7bcaefa..461179a0982 100644
--- a/src/gallium/drivers/radeonsi/si_state.c
+++ b/src/gallium/drivers/radeonsi/si_state.c
@@ -1836,12 +1836,12 @@ static void si_set_active_query_state(struct pipe_context *ctx, bool enable)
 
 void si_save_qbo_state(struct si_context *sctx, struct si_qbo_state *st)
 {
-   si_get_pipe_constant_buffer(sctx, PIPE_SHADER_COMPUTE, 0, &st->saved_const0);
+   si_get_pipe_constant_buffer(sctx, MESA_SHADER_COMPUTE, 0, &st->saved_const0);
 }
 
 void si_restore_qbo_state(struct si_context *sctx, struct si_qbo_state *st)
 {
-   sctx->b.set_constant_buffer(&sctx->b, PIPE_SHADER_COMPUTE, 0, true, &st->saved_const0);
+   sctx->b.set_constant_buffer(&sctx->b, MESA_SHADER_COMPUTE, 0, true, &st->saved_const0);
 }
 
 static void si_emit_db_render_state(struct si_context *sctx, unsigned index)
diff --git a/src/gallium/drivers/radeonsi/si_state.h b/src/gallium/drivers/radeonsi/si_state.h
index 48abfddd8e5..615a07df030 100644
--- a/src/gallium/drivers/radeonsi/si_state.h
+++ b/src/gallium/drivers/radeonsi/si_state.h
@@ -16,7 +16,7 @@ extern "C" {
 #endif
 
 #define SI_NUM_GRAPHICS_SHADERS (MESA_SHADER_FRAGMENT + 1)
-#define SI_NUM_SHADERS          (PIPE_SHADER_COMPUTE + 1)
+#define SI_NUM_SHADERS          (MESA_SHADER_COMPUTE + 1)
 
 #define SI_NUM_VERTEX_BUFFERS SI_MAX_ATTRIBS
 #define SI_NUM_SAMPLERS       32 /* OpenGL textures units per shader */
@@ -497,7 +497,7 @@ enum
 
 #define SI_DESCS_INTERNAL      0
 #define SI_DESCS_FIRST_SHADER  1
-#define SI_DESCS_FIRST_COMPUTE (SI_DESCS_FIRST_SHADER + PIPE_SHADER_COMPUTE * SI_NUM_SHADER_DESCS)
+#define SI_DESCS_FIRST_COMPUTE (SI_DESCS_FIRST_SHADER + MESA_SHADER_COMPUTE * SI_NUM_SHADER_DESCS)
 #define SI_NUM_DESCS           (SI_DESCS_FIRST_SHADER + SI_NUM_SHADERS * SI_NUM_SHADER_DESCS)
 
 #define SI_DESCS_SHADER_MASK(name)                                                                 \
diff --git a/src/gallium/drivers/softpipe/sp_compute.c b/src/gallium/drivers/softpipe/sp_compute.c
index 15f0f3ede58..e6f63fdf463 100644
--- a/src/gallium/drivers/softpipe/sp_compute.c
+++ b/src/gallium/drivers/softpipe/sp_compute.c
@@ -199,7 +199,7 @@ softpipe_launch_grid(struct pipe_context *context,
    for (local_z = 0; local_z < bdepth; local_z++) {
       for (local_y = 0; local_y < bheight; local_y++) {
          for (local_x = 0; local_x < bwidth; local_x += TGSI_QUAD_SIZE) {
-            machines[idx] = tgsi_exec_machine_create(PIPE_SHADER_COMPUTE);
+            machines[idx] = tgsi_exec_machine_create(MESA_SHADER_COMPUTE);
 
             machines[idx]->LocalMem = local_mem;
             machines[idx]->LocalMemSize = shared_mem_size;
@@ -208,11 +208,11 @@ softpipe_launch_grid(struct pipe_context *context,
                        local_x, local_y, local_z,
                        grid_size[0], grid_size[1], grid_size[2],
                        bwidth, bheight, bdepth,
-                       (struct tgsi_sampler *)softpipe->tgsi.sampler[PIPE_SHADER_COMPUTE],
-                       (struct tgsi_image *)softpipe->tgsi.image[PIPE_SHADER_COMPUTE],
-                       (struct tgsi_buffer *)softpipe->tgsi.buffer[PIPE_SHADER_COMPUTE]);
+                       (struct tgsi_sampler *)softpipe->tgsi.sampler[MESA_SHADER_COMPUTE],
+                       (struct tgsi_image *)softpipe->tgsi.image[MESA_SHADER_COMPUTE],
+                       (struct tgsi_buffer *)softpipe->tgsi.buffer[MESA_SHADER_COMPUTE]);
             tgsi_exec_set_constant_buffers(machines[idx], PIPE_MAX_CONSTANT_BUFFERS,
-                                           softpipe->mapped_constants[PIPE_SHADER_COMPUTE]);
+                                           softpipe->mapped_constants[MESA_SHADER_COMPUTE]);
             idx++;
          }
       }
diff --git a/src/gallium/drivers/softpipe/sp_screen.c b/src/gallium/drivers/softpipe/sp_screen.c
index 6e85cd23b34..aeec2caf0a3 100644
--- a/src/gallium/drivers/softpipe/sp_screen.c
+++ b/src/gallium/drivers/softpipe/sp_screen.c
@@ -200,7 +200,7 @@ softpipe_is_format_supported( struct pipe_screen *screen,
 static void
 softpipe_init_shader_caps(struct softpipe_screen *sp_screen)
 {
-   for (unsigned i = 0; i <= PIPE_SHADER_COMPUTE; i++) {
+   for (unsigned i = 0; i <= MESA_SHADER_COMPUTE; i++) {
       struct pipe_shader_caps *caps =
          (struct pipe_shader_caps *)&sp_screen->base.shader_caps[i];
 
@@ -213,7 +213,7 @@ softpipe_init_shader_caps(struct softpipe_screen *sp_screen)
          }
          FALLTHROUGH;
       case MESA_SHADER_FRAGMENT:
-      case PIPE_SHADER_COMPUTE:
+      case MESA_SHADER_COMPUTE:
          tgsi_exec_init_shader_caps(caps);
          break;
       default:
diff --git a/src/gallium/drivers/softpipe/sp_state_derived.c b/src/gallium/drivers/softpipe/sp_state_derived.c
index 6d539f110de..195b9940488 100644
--- a/src/gallium/drivers/softpipe/sp_state_derived.c
+++ b/src/gallium/drivers/softpipe/sp_state_derived.c
@@ -294,7 +294,7 @@ set_shader_sampler(struct softpipe_context *softpipe,
 void
 softpipe_update_compute_samplers(struct softpipe_context *softpipe)
 {
-   set_shader_sampler(softpipe, PIPE_SHADER_COMPUTE, softpipe->cs->max_sampler);
+   set_shader_sampler(softpipe, MESA_SHADER_COMPUTE, softpipe->cs->max_sampler);
 }
 
 static void
diff --git a/src/gallium/drivers/svga/svga_draw.c b/src/gallium/drivers/svga/svga_draw.c
index 778b67732f9..c4640423b25 100644
--- a/src/gallium/drivers/svga/svga_draw.c
+++ b/src/gallium/drivers/svga/svga_draw.c
@@ -339,11 +339,11 @@ svga_validate_sampler_resources(struct svga_context *svga,
 
    if (pipe_type == SVGA_PIPE_GRAPHICS) {
       first_shader = MESA_SHADER_VERTEX;
-      last_shader = PIPE_SHADER_COMPUTE;
+      last_shader = MESA_SHADER_COMPUTE;
    }
    else {
       assert(svga_have_gl43(svga));
-      first_shader = PIPE_SHADER_COMPUTE;
+      first_shader = MESA_SHADER_COMPUTE;
       last_shader = first_shader+1;
    }
 
@@ -422,11 +422,11 @@ svga_validate_constant_buffers(struct svga_context *svga,
 
    if (pipe_type == SVGA_PIPE_GRAPHICS) {
       first_shader = MESA_SHADER_VERTEX;
-      last_shader = PIPE_SHADER_COMPUTE;
+      last_shader = MESA_SHADER_COMPUTE;
    }
    else {
       assert(svga_have_gl43(svga));
-      first_shader = PIPE_SHADER_COMPUTE;
+      first_shader = MESA_SHADER_COMPUTE;
       last_shader = first_shader + 1;
    }
 
@@ -524,10 +524,10 @@ svga_validate_image_views(struct svga_context *svga,
 
    if (pipe_type == SVGA_PIPE_GRAPHICS) {
       first_shader = MESA_SHADER_VERTEX;
-      last_shader = PIPE_SHADER_COMPUTE;
+      last_shader = MESA_SHADER_COMPUTE;
    }
    else {
-      first_shader = PIPE_SHADER_COMPUTE;
+      first_shader = MESA_SHADER_COMPUTE;
       last_shader = first_shader + 1;
    }
 
@@ -563,10 +563,10 @@ svga_validate_shader_buffers(struct svga_context *svga,
 
    if (pipe_type == SVGA_PIPE_GRAPHICS) {
       first_shader = MESA_SHADER_VERTEX;
-      last_shader = PIPE_SHADER_COMPUTE;
+      last_shader = MESA_SHADER_COMPUTE;
    }
    else {
-      first_shader = PIPE_SHADER_COMPUTE;
+      first_shader = MESA_SHADER_COMPUTE;
       last_shader = first_shader + 1;
    }
 
diff --git a/src/gallium/drivers/svga/svga_pipe_constants.c b/src/gallium/drivers/svga/svga_pipe_constants.c
index 4a4d1ae939b..910d74c77a9 100644
--- a/src/gallium/drivers/svga/svga_pipe_constants.c
+++ b/src/gallium/drivers/svga/svga_pipe_constants.c
@@ -76,7 +76,7 @@ svga_set_constant_buffer(struct pipe_context *pipe,
          svga->dirty |= SVGA_NEW_TCS_CONSTS;
       else if (shader == MESA_SHADER_TESS_EVAL)
          svga->dirty |= SVGA_NEW_TES_CONSTS;
-      else if (shader == PIPE_SHADER_COMPUTE)
+      else if (shader == MESA_SHADER_COMPUTE)
          svga->dirty |= SVGA_NEW_CS_CONSTS;
    } else {
       if (shader == MESA_SHADER_FRAGMENT)
@@ -89,7 +89,7 @@ svga_set_constant_buffer(struct pipe_context *pipe,
          svga->dirty |= SVGA_NEW_TCS_CONST_BUFFER;
       else if (shader == MESA_SHADER_TESS_EVAL)
          svga->dirty |= SVGA_NEW_TES_CONST_BUFFER;
-      else if (shader == PIPE_SHADER_COMPUTE)
+      else if (shader == MESA_SHADER_COMPUTE)
          svga->dirty |= SVGA_NEW_CS_CONST_BUFFER;
 
       /* update bitmask of dirty const buffers */
diff --git a/src/gallium/drivers/svga/svga_pipe_cs.c b/src/gallium/drivers/svga/svga_pipe_cs.c
index 996cd6e16ee..2e8037e7392 100644
--- a/src/gallium/drivers/svga/svga_pipe_cs.c
+++ b/src/gallium/drivers/svga/svga_pipe_cs.c
@@ -48,7 +48,7 @@ svga_create_compute_state(struct pipe_context *pipe,
    struct svga_shader *shader = &cs->base;
    shader->id = svga->debug.shader_id++;
    shader->type = PIPE_SHADER_IR_TGSI;
-   shader->stage = PIPE_SHADER_COMPUTE;
+   shader->stage = MESA_SHADER_COMPUTE;
 
    /* Collect shader basic info */
    svga_tgsi_scan_shader(&cs->base);
@@ -73,7 +73,7 @@ svga_bind_compute_state(struct pipe_context *pipe, void *shader)
    svga->dirty |= SVGA_NEW_CS;
 
    /* Check if the shader uses samplers */
-   svga_set_curr_shader_use_samplers_flag(svga, PIPE_SHADER_COMPUTE,
+   svga_set_curr_shader_use_samplers_flag(svga, MESA_SHADER_COMPUTE,
                                           svga_shader_use_samplers(&cs->base));
 }
 
diff --git a/src/gallium/drivers/svga/svga_pipe_sampler.c b/src/gallium/drivers/svga/svga_pipe_sampler.c
index 8efa8debdc2..15fb78e53ab 100644
--- a/src/gallium/drivers/svga/svga_pipe_sampler.c
+++ b/src/gallium/drivers/svga/svga_pipe_sampler.c
@@ -511,7 +511,7 @@ svga_cleanup_sampler_state(struct svga_context *svga)
 {
    enum pipe_shader_type shader;
 
-   for (shader = 0; shader <= PIPE_SHADER_COMPUTE; shader++) {
+   for (shader = 0; shader <= MESA_SHADER_COMPUTE; shader++) {
       unsigned i;
 
       for (i = 0; i < svga->state.hw_draw.num_sampler_views[shader]; i++) {
diff --git a/src/gallium/drivers/svga/svga_screen.c b/src/gallium/drivers/svga/svga_screen.c
index 31185d35496..7cf97436a67 100644
--- a/src/gallium/drivers/svga/svga_screen.c
+++ b/src/gallium/drivers/svga/svga_screen.c
@@ -256,7 +256,7 @@ vgpu10_init_shader_caps(struct svga_screen *svgascreen)
 
    assert(sws->have_vgpu10);
 
-    for (unsigned i = 0; i <= PIPE_SHADER_COMPUTE; i++) {
+    for (unsigned i = 0; i <= MESA_SHADER_COMPUTE; i++) {
        struct pipe_shader_caps *caps =
           (struct pipe_shader_caps *)&svgascreen->screen.shader_caps[i];
 
@@ -266,7 +266,7 @@ vgpu10_init_shader_caps(struct svga_screen *svgascreen)
           if (!sws->have_sm5)
              continue;
           break;
-       case PIPE_SHADER_COMPUTE:
+       case MESA_SHADER_COMPUTE:
           if (!sws->have_gl43)
              continue;
           break;
diff --git a/src/gallium/drivers/svga/svga_shader.c b/src/gallium/drivers/svga/svga_shader.c
index c90515d0bca..19fbe3b0734 100644
--- a/src/gallium/drivers/svga/svga_shader.c
+++ b/src/gallium/drivers/svga/svga_shader.c
@@ -473,7 +473,7 @@ svga_init_shader_key_common(const struct svga_context *svga,
          /* Save the uavSpliceIndex which is the index used for the first uav
           * in the draw pipeline. For compute, uavSpliceIndex is always 0.
           */
-         if (shader_type != PIPE_SHADER_COMPUTE)
+         if (shader_type != MESA_SHADER_COMPUTE)
             key->uav_splice_index = svga->state.hw_draw.uavSpliceIndex;
 
          unsigned uav_splice_index = key->uav_splice_index;
@@ -791,7 +791,7 @@ svga_new_shader_variant(struct svga_context *svga, enum pipe_shader_type type)
    case MESA_SHADER_TESS_CTRL:
       variant = CALLOC(1, sizeof(struct svga_tcs_variant));
       break;
-   case PIPE_SHADER_COMPUTE:
+   case MESA_SHADER_COMPUTE:
       variant = CALLOC(1, sizeof(struct svga_cs_variant));
       break;
    default:
diff --git a/src/gallium/drivers/svga/svga_shader.h b/src/gallium/drivers/svga/svga_shader.h
index 281896b985f..6eafa0ce1c4 100644
--- a/src/gallium/drivers/svga/svga_shader.h
+++ b/src/gallium/drivers/svga/svga_shader.h
@@ -542,7 +542,7 @@ svga_shader_type(enum pipe_shader_type shader)
       return SVGA3D_SHADERTYPE_HS;
    case MESA_SHADER_TESS_EVAL:
       return SVGA3D_SHADERTYPE_DS;
-   case PIPE_SHADER_COMPUTE:
+   case MESA_SHADER_COMPUTE:
       return SVGA3D_SHADERTYPE_CS;
    default:
       assert(!"Invalid shader type");
diff --git a/src/gallium/drivers/svga/svga_state_constants.c b/src/gallium/drivers/svga/svga_state_constants.c
index efd32668986..773fddc03fd 100644
--- a/src/gallium/drivers/svga/svga_state_constants.c
+++ b/src/gallium/drivers/svga/svga_state_constants.c
@@ -385,7 +385,7 @@ svga_get_extra_cs_constants(struct svga_context *svga, float *dest)
 
    /* common constants */
    count += svga_get_extra_constants_common(svga, variant,
-                                            PIPE_SHADER_COMPUTE,
+                                            MESA_SHADER_COMPUTE,
                                             dest);
 
    assert(count <= MAX_EXTRA_CONSTS);
@@ -888,7 +888,7 @@ emit_consts_vgpu10(struct svga_context *svga, enum pipe_shader_type shader)
           shader == MESA_SHADER_FRAGMENT ||
           shader == MESA_SHADER_TESS_CTRL ||
           shader == MESA_SHADER_TESS_EVAL ||
-          shader == PIPE_SHADER_COMPUTE);
+          shader == MESA_SHADER_COMPUTE);
 
    cbuf = &svga->curr.constbufs[shader][0];
 
@@ -913,7 +913,7 @@ emit_consts_vgpu10(struct svga_context *svga, enum pipe_shader_type shader)
       variant = svga->state.hw_draw.tes;
       extra_count = svga_get_extra_tes_constants(svga, (float *) extras);
       break;
-   case PIPE_SHADER_COMPUTE:
+   case MESA_SHADER_COMPUTE:
       variant = svga->state.hw_draw.cs;
       extra_count = svga_get_extra_cs_constants(svga, (float *) extras);
       break;
@@ -1369,7 +1369,7 @@ emit_cs_consts(struct svga_context *svga, uint64_t dirty)
       return PIPE_OK;
 
    /* SVGA_NEW_CS_CONST_BUFFER */
-   ret = emit_consts_vgpu10(svga, PIPE_SHADER_COMPUTE);
+   ret = emit_consts_vgpu10(svga, MESA_SHADER_COMPUTE);
 
    return ret;
 }
@@ -1389,7 +1389,7 @@ emit_cs_constbuf(struct svga_context *svga, uint64_t dirty)
    /* SVGA_NEW_CS_CONSTBUF
     */
    assert(svga_have_vgpu10(svga));
-   ret = emit_constbuf_vgpu10(svga, PIPE_SHADER_COMPUTE);
+   ret = emit_constbuf_vgpu10(svga, MESA_SHADER_COMPUTE);
 
    return ret;
 }
@@ -1458,7 +1458,7 @@ update_rawbuf(struct svga_context *svga, uint64 dirty)
    };
 
    for (enum pipe_shader_type shader = MESA_SHADER_VERTEX;
-        shader < PIPE_SHADER_COMPUTE; shader++) {
+        shader < MESA_SHADER_COMPUTE; shader++) {
       unsigned rawbuf_mask = svga->state.raw_constbufs[shader];
       unsigned rawbuf_sbuf_mask = svga->state.raw_shaderbufs[shader];
 
@@ -1496,15 +1496,15 @@ struct svga_tracked_state svga_need_rawbuf_srv =
 static enum pipe_error
 update_cs_rawbuf(struct svga_context *svga, uint64 dirty)
 {
-   unsigned rawbuf_mask = svga->state.raw_constbufs[PIPE_SHADER_COMPUTE];
+   unsigned rawbuf_mask = svga->state.raw_constbufs[MESA_SHADER_COMPUTE];
 
-   update_rawbuf_mask(svga, PIPE_SHADER_COMPUTE);
+   update_rawbuf_mask(svga, MESA_SHADER_COMPUTE);
 
    /* if the rawbuf state is different for the shader stage,
     * send SVGA_NEW_RAW_BUFFER to trigger a new shader
     * variant to use srv for ubo access.
     */
-   if (svga->state.raw_constbufs[PIPE_SHADER_COMPUTE] != rawbuf_mask)
+   if (svga->state.raw_constbufs[MESA_SHADER_COMPUTE] != rawbuf_mask)
       svga->dirty |= SVGA_NEW_CS_RAW_BUFFER;
 
    return PIPE_OK;
diff --git a/src/gallium/drivers/svga/svga_state_cs.c b/src/gallium/drivers/svga/svga_state_cs.c
index a7e62e7ec9b..7427210e8be 100644
--- a/src/gallium/drivers/svga/svga_state_cs.c
+++ b/src/gallium/drivers/svga/svga_state_cs.c
@@ -27,7 +27,7 @@ make_cs_key(struct svga_context *svga,
 
    memset(key, 0, sizeof *key);
 
-   svga_init_shader_key_common(svga, PIPE_SHADER_COMPUTE, &cs->base, key);
+   svga_init_shader_key_common(svga, MESA_SHADER_COMPUTE, &cs->base, key);
 
    key->cs.grid_size[0] = svga->curr.grid_info.size[0];
    key->cs.grid_size[1] = svga->curr.grid_info.size[1];
diff --git a/src/gallium/drivers/svga/svga_state_sampler.c b/src/gallium/drivers/svga/svga_state_sampler.c
index 8cc9b9be589..c70aa39b1e6 100644
--- a/src/gallium/drivers/svga/svga_state_sampler.c
+++ b/src/gallium/drivers/svga/svga_state_sampler.c
@@ -231,7 +231,7 @@ update_sampler_resources(struct svga_context *svga, uint64_t dirty)
 
    assert(svga_have_vgpu10(svga));
 
-   for (shader = MESA_SHADER_VERTEX; shader < PIPE_SHADER_COMPUTE; shader++) {
+   for (shader = MESA_SHADER_VERTEX; shader < MESA_SHADER_COMPUTE; shader++) {
       SVGA3dShaderResourceViewId ids[PIPE_MAX_SAMPLERS];
       struct svga_winsys_surface *surfaces[PIPE_MAX_SAMPLERS];
       struct pipe_sampler_view *sampler_views[PIPE_MAX_SAMPLERS];
@@ -382,7 +382,7 @@ update_samplers(struct svga_context *svga, uint64_t dirty )
 
    assert(svga_have_vgpu10(svga));
 
-   for (shader = MESA_SHADER_VERTEX; shader < PIPE_SHADER_COMPUTE; shader++) {
+   for (shader = MESA_SHADER_VERTEX; shader < MESA_SHADER_COMPUTE; shader++) {
       const unsigned count = svga->curr.num_samplers[shader];
       SVGA3dSamplerId ids[PIPE_MAX_SAMPLERS*2];
       unsigned i;
@@ -527,7 +527,7 @@ static enum pipe_error
 update_cs_sampler_resources(struct svga_context *svga, uint64_t dirty)
 {
    enum pipe_error ret = PIPE_OK;
-   enum pipe_shader_type shader = PIPE_SHADER_COMPUTE;
+   enum pipe_shader_type shader = MESA_SHADER_COMPUTE;
 
    assert(svga_have_sm5(svga));
 
@@ -649,7 +649,7 @@ static enum pipe_error
 update_cs_samplers(struct svga_context *svga, uint64_t dirty )
 {
    enum pipe_error ret = PIPE_OK;
-   enum pipe_shader_type shader = PIPE_SHADER_COMPUTE;
+   enum pipe_shader_type shader = MESA_SHADER_COMPUTE;
 
    assert(svga_have_sm5(svga));
 
diff --git a/src/gallium/drivers/svga/svga_state_uav.c b/src/gallium/drivers/svga/svga_state_uav.c
index 64dafe5306a..d18819fa4f3 100644
--- a/src/gallium/drivers/svga/svga_state_uav.c
+++ b/src/gallium/drivers/svga/svga_state_uav.c
@@ -344,9 +344,9 @@ svga_create_uav_list(struct svga_context *svga,
 
    if (pipe_type == SVGA_PIPE_GRAPHICS) {
       first_shader = MESA_SHADER_VERTEX;
-      last_shader = PIPE_SHADER_COMPUTE;
+      last_shader = MESA_SHADER_COMPUTE;
    } else {
-      first_shader = PIPE_SHADER_COMPUTE;
+      first_shader = MESA_SHADER_COMPUTE;
       last_shader = first_shader + 1;
    }
 
@@ -594,9 +594,9 @@ svga_save_uav_state(struct svga_context *svga,
 
    if (pipe_type == SVGA_PIPE_GRAPHICS) {
       first_shader = MESA_SHADER_VERTEX;
-      last_shader = PIPE_SHADER_COMPUTE;
+      last_shader = MESA_SHADER_COMPUTE;
    } else {
-      first_shader = PIPE_SHADER_COMPUTE;
+      first_shader = MESA_SHADER_COMPUTE;
       last_shader = first_shader + 1;
    }
 
@@ -703,7 +703,7 @@ need_to_set_uav(struct svga_context *svga,
 
    /* If image views are different */
    for (enum pipe_shader_type shader = MESA_SHADER_VERTEX;
-        shader < PIPE_SHADER_COMPUTE; shader++) {
+        shader < MESA_SHADER_COMPUTE; shader++) {
       unsigned num_image_views = svga->curr.num_image_views[shader];
       if ((num_image_views != svga->state.hw_draw.num_image_views[shader]) ||
           memcmp(svga->state.hw_draw.image_views[shader],
@@ -811,7 +811,7 @@ need_to_set_cs_uav(struct svga_context *svga,
                    SVGA3dUAViewId *uaViewIds,
                    struct svga_winsys_surface **uaViews)
 {
-   enum pipe_shader_type shader = PIPE_SHADER_COMPUTE;
+   enum pipe_shader_type shader = MESA_SHADER_COMPUTE;
 
    if (svga->state.hw_draw.num_cs_uavs != num_uavs)
       return true;
diff --git a/src/gallium/drivers/svga/svga_surface.c b/src/gallium/drivers/svga/svga_surface.c
index c86fecc1c99..d100a8dcccc 100644
--- a/src/gallium/drivers/svga/svga_surface.c
+++ b/src/gallium/drivers/svga/svga_surface.c
@@ -522,7 +522,7 @@ svga_validate_surface_view(struct svga_context *svga, struct svga_surface *s)
     * associated resource. We will then use the cloned surface view for
     * render target.
     */
-   for (shader = MESA_SHADER_VERTEX; shader <= PIPE_SHADER_COMPUTE; shader++) {
+   for (shader = MESA_SHADER_VERTEX; shader <= MESA_SHADER_COMPUTE; shader++) {
       if (svga_check_sampler_view_resource_collision(svga, s->handle, shader)) {
          SVGA_DBG(DEBUG_VIEWS,
                   "same resource used in shaderResource and renderTarget 0x%x\n",
diff --git a/src/gallium/drivers/svga/svga_tgsi_vgpu10.c b/src/gallium/drivers/svga/svga_tgsi_vgpu10.c
index 54b2a790df5..0a62793e325 100644
--- a/src/gallium/drivers/svga/svga_tgsi_vgpu10.c
+++ b/src/gallium/drivers/svga/svga_tgsi_vgpu10.c
@@ -825,7 +825,7 @@ translate_shader_type(unsigned type)
       return VGPU10_HULL_SHADER;
    case MESA_SHADER_TESS_EVAL:
       return VGPU10_DOMAIN_SHADER;
-   case PIPE_SHADER_COMPUTE:
+   case MESA_SHADER_COMPUTE:
       return VGPU10_COMPUTE_SHADER;
    default:
       assert(!"Unexpected shader type");
@@ -1823,7 +1823,7 @@ emit_src_register(struct svga_shader_emitter_v10 *emit,
          }
       }
    }
-   else if (emit->unit == PIPE_SHADER_COMPUTE) {
+   else if (emit->unit == MESA_SHADER_COMPUTE) {
       if (file == TGSI_FILE_SYSTEM_VALUE) {
          if (index == emit->cs.thread_id_index) {
             operand0.numComponents = VGPU10_OPERAND_4_COMPONENT;
@@ -4228,7 +4228,7 @@ emit_vertex_output_declaration(struct svga_shader_emitter_v10 *emit,
    unsigned final_mask = VGPU10_OPERAND_4_COMPONENT_MASK_ALL;
 
    assert(emit->unit != MESA_SHADER_FRAGMENT &&
-          emit->unit != PIPE_SHADER_COMPUTE);
+          emit->unit != MESA_SHADER_COMPUTE);
 
    switch (semantic_name) {
    case TGSI_SEMANTIC_POSITION:
@@ -4865,7 +4865,7 @@ emit_system_value_declaration(struct svga_shader_emitter_v10 *emit,
       }
       break;
    case TGSI_SEMANTIC_THREAD_ID:
-      assert(emit->unit >= PIPE_SHADER_COMPUTE);
+      assert(emit->unit >= MESA_SHADER_COMPUTE);
       assert(emit->version >= 50);
       emit->cs.thread_id_index = index;
       emit_input_declaration(emit, VGPU10_OPCODE_DCL_INPUT,
@@ -4880,7 +4880,7 @@ emit_system_value_declaration(struct svga_shader_emitter_v10 *emit,
                              map_tgsi_semantic_to_sgn_name(semantic_name));
       break;
    case TGSI_SEMANTIC_BLOCK_ID:
-      assert(emit->unit >= PIPE_SHADER_COMPUTE);
+      assert(emit->unit >= MESA_SHADER_COMPUTE);
       assert(emit->version >= 50);
       emit->cs.block_id_index = index;
       emit_input_declaration(emit, VGPU10_OPCODE_DCL_INPUT,
@@ -4895,7 +4895,7 @@ emit_system_value_declaration(struct svga_shader_emitter_v10 *emit,
                              map_tgsi_semantic_to_sgn_name(semantic_name));
       break;
    case TGSI_SEMANTIC_GRID_SIZE:
-      assert(emit->unit == PIPE_SHADER_COMPUTE);
+      assert(emit->unit == MESA_SHADER_COMPUTE);
       assert(emit->version >= 50);
       emit->cs.grid_size.tgsi_index = index;
       break;
@@ -5037,7 +5037,7 @@ emit_vgpu10_declaration(struct svga_shader_emitter_v10 *emit,
 
    case TGSI_FILE_MEMORY:
       /* Record memory has been used. */
-      if (emit->unit == PIPE_SHADER_COMPUTE &&
+      if (emit->unit == MESA_SHADER_COMPUTE &&
           decl->Declaration.MemType == TGSI_MEMORY_TYPE_SHARED) {
          emit->cs.shared_memory_declared = true;
       }
@@ -5509,7 +5509,7 @@ emit_input_declarations(struct svga_shader_emitter_v10 *emit)
    case MESA_SHADER_TESS_EVAL:
       emit_tes_input_declarations(emit);
       break;
-   case PIPE_SHADER_COMPUTE:
+   case MESA_SHADER_COMPUTE:
       //XXX emit_cs_input_declarations(emit);
       break;
    default:
@@ -5549,7 +5549,7 @@ emit_output_declarations(struct svga_shader_emitter_v10 *emit)
    case MESA_SHADER_TESS_EVAL:
       emit_tes_output_declarations(emit);
       break;
-   case PIPE_SHADER_COMPUTE:
+   case MESA_SHADER_COMPUTE:
       //XXX emit_cs_output_declarations(emit);
       break;
    default:
@@ -5991,7 +5991,7 @@ emit_constant_declaration(struct svga_shader_emitter_v10 *emit)
    if (emit->key.clip_plane_enable) {
       unsigned n = util_bitcount(emit->key.clip_plane_enable);
       assert(emit->unit != MESA_SHADER_FRAGMENT &&
-             emit->unit != PIPE_SHADER_COMPUTE);
+             emit->unit != MESA_SHADER_COMPUTE);
       for (i = 0; i < n; i++) {
          emit->clip_plane_const[i] = total_consts++;
       }
@@ -10842,7 +10842,7 @@ emit_barrier(struct svga_shader_emitter_v10 *emit,
                          "barrier instruction is not supported in tessellation control shader\n");
       return true;
    }
-   else if (emit->unit == PIPE_SHADER_COMPUTE) {
+   else if (emit->unit == MESA_SHADER_COMPUTE) {
       if (emit->cs.shared_memory_declared)
          token0.syncThreadGroupShared = 1;
 
@@ -10879,7 +10879,7 @@ emit_memory_barrier(struct svga_shader_emitter_v10 *emit,
    token0.value = 0;
    token0.opcodeType = VGPU10_OPCODE_SYNC;
 
-   if (emit->unit == PIPE_SHADER_COMPUTE) {
+   if (emit->unit == MESA_SHADER_COMPUTE) {
 
       /* For compute shader, issue sync opcode with different options
        * depending on the memory barrier type.
@@ -12156,7 +12156,7 @@ emit_pre_helpers(struct svga_shader_emitter_v10 *emit)
    else if (emit->unit == MESA_SHADER_TESS_EVAL) {
       emit_domain_shader_declarations(emit);
    }
-   else if (emit->unit == PIPE_SHADER_COMPUTE) {
+   else if (emit->unit == MESA_SHADER_COMPUTE) {
       emit_compute_shader_declarations(emit);
    }
 
@@ -12196,7 +12196,7 @@ emit_pre_helpers(struct svga_shader_emitter_v10 *emit)
    }
 
    if (emit->unit != MESA_SHADER_FRAGMENT &&
-       emit->unit != PIPE_SHADER_COMPUTE) {
+       emit->unit != MESA_SHADER_COMPUTE) {
       /*
        * Declare clip distance output registers for ClipVertex or
        * user defined planes
@@ -12204,7 +12204,7 @@ emit_pre_helpers(struct svga_shader_emitter_v10 *emit)
       emit_clip_distance_declarations(emit);
    }
 
-   if (emit->unit == PIPE_SHADER_COMPUTE) {
+   if (emit->unit == MESA_SHADER_COMPUTE) {
       emit_memory_declarations(emit);
 
       if (emit->cs.grid_size.tgsi_index != INVALID_INDEX) {
@@ -12925,7 +12925,7 @@ svga_tgsi_vgpu10_translate(struct svga_context *svga,
           unit == MESA_SHADER_FRAGMENT ||
           unit == MESA_SHADER_TESS_CTRL ||
           unit == MESA_SHADER_TESS_EVAL ||
-          unit == PIPE_SHADER_COMPUTE);
+          unit == MESA_SHADER_COMPUTE);
 
    /* These two flags cannot be used together */
    assert(key->vs.need_prescale + key->vs.undo_viewport <= 1);
diff --git a/src/gallium/drivers/v3d/v3d_context.c b/src/gallium/drivers/v3d/v3d_context.c
index 2aa75a0e757..d747aa0b13b 100644
--- a/src/gallium/drivers/v3d/v3d_context.c
+++ b/src/gallium/drivers/v3d/v3d_context.c
@@ -248,7 +248,7 @@ v3d_flag_dirty_sampler_state(struct v3d_context *v3d,
         case MESA_SHADER_FRAGMENT:
                 v3d->dirty |= V3D_DIRTY_FRAGTEX;
                 break;
-        case PIPE_SHADER_COMPUTE:
+        case MESA_SHADER_COMPUTE:
                 v3d->dirty |= V3D_DIRTY_COMPTEX;
                 break;
         default:
diff --git a/src/gallium/drivers/v3d/v3d_program.c b/src/gallium/drivers/v3d/v3d_program.c
index 4aad99150ea..ef9a8344927 100644
--- a/src/gallium/drivers/v3d/v3d_program.c
+++ b/src/gallium/drivers/v3d/v3d_program.c
@@ -942,7 +942,7 @@ v3d_update_compiled_cs(struct v3d_context *v3d)
         }
 
         memset(key, 0, sizeof(*key));
-        v3d_setup_shared_key(v3d, key, &v3d->tex[PIPE_SHADER_COMPUTE]);
+        v3d_setup_shared_key(v3d, key, &v3d->tex[MESA_SHADER_COMPUTE]);
 
         struct v3d_compiled_shader *cs =
                 v3d_get_compiled_shader(v3d, key, sizeof(*key),
diff --git a/src/gallium/drivers/v3d/v3d_screen.c b/src/gallium/drivers/v3d/v3d_screen.c
index 5bb50705be4..fc26cdb89ea 100644
--- a/src/gallium/drivers/v3d/v3d_screen.c
+++ b/src/gallium/drivers/v3d/v3d_screen.c
@@ -119,7 +119,7 @@ v3d_has_feature(struct v3d_screen *screen, enum drm_v3d_param feature)
 static void
 v3d_init_shader_caps(struct v3d_screen *screen)
 {
-        for (unsigned i = 0; i <= PIPE_SHADER_COMPUTE; i++) {
+        for (unsigned i = 0; i <= MESA_SHADER_COMPUTE; i++) {
                 struct pipe_shader_caps *caps =
                         (struct pipe_shader_caps *)&screen->base.shader_caps[i];
 
@@ -128,7 +128,7 @@ v3d_init_shader_caps(struct v3d_screen *screen)
                 case MESA_SHADER_FRAGMENT:
                 case MESA_SHADER_GEOMETRY:
                         break;
-                case PIPE_SHADER_COMPUTE:
+                case MESA_SHADER_COMPUTE:
                         if (!screen->has_csd)
                                 continue;
                         break;
diff --git a/src/gallium/drivers/v3d/v3dx_draw.c b/src/gallium/drivers/v3d/v3dx_draw.c
index 985068114de..139f15e7fc4 100644
--- a/src/gallium/drivers/v3d/v3dx_draw.c
+++ b/src/gallium/drivers/v3d/v3dx_draw.c
@@ -147,7 +147,7 @@ v3d_predraw_check_stage_inputs(struct pipe_context *pctx,
 
                 v3d_flush_jobs_writing_resource(v3d, view->texture,
                                                 V3D_FLUSH_NOT_CURRENT_JOB,
-                                                s == PIPE_SHADER_COMPUTE);
+                                                s == MESA_SHADER_COMPUTE);
         }
 
         /* Flush writes to UBOs. */
@@ -157,7 +157,7 @@ v3d_predraw_check_stage_inputs(struct pipe_context *pctx,
                 if (cb->buffer) {
                         v3d_flush_jobs_writing_resource(v3d, cb->buffer,
                                                         V3D_FLUSH_DEFAULT,
-                                                        s == PIPE_SHADER_COMPUTE);
+                                                        s == MESA_SHADER_COMPUTE);
                 }
         }
 
@@ -168,7 +168,7 @@ v3d_predraw_check_stage_inputs(struct pipe_context *pctx,
                 if (sb->buffer) {
                         v3d_flush_jobs_reading_resource(v3d, sb->buffer,
                                                         V3D_FLUSH_NOT_CURRENT_JOB,
-                                                        s == PIPE_SHADER_COMPUTE);
+                                                        s == MESA_SHADER_COMPUTE);
                 }
         }
 
@@ -179,7 +179,7 @@ v3d_predraw_check_stage_inputs(struct pipe_context *pctx,
 
                 v3d_flush_jobs_reading_resource(v3d, view->base.resource,
                                                 V3D_FLUSH_NOT_CURRENT_JOB,
-                                                s == PIPE_SHADER_COMPUTE);
+                                                s == MESA_SHADER_COMPUTE);
         }
 
         /* Flush writes to our vertex buffers (i.e. from transform feedback) */
@@ -312,7 +312,7 @@ v3d_emit_wait_for_tf_if_needed(struct v3d_context *v3d, struct v3d_job *job)
 
         set_foreach(job->tf_write_prscs, entry) {
                 struct pipe_resource *prsc = (struct pipe_resource *)entry->key;
-                for (int s = 0; s < PIPE_SHADER_COMPUTE; s++) {
+                for (int s = 0; s < MESA_SHADER_COMPUTE; s++) {
                         /* Fragment shaders can only start executing after all
                          * binning (and thus TF) is complete.
                          *
@@ -1145,7 +1145,7 @@ v3d_draw_vbo(struct pipe_context *pctx, const struct pipe_draw_info *info,
         /* Before setting up the draw, flush anything writing to the resources
          * that we read from or reading from resources we write to.
          */
-        for (int s = 0; s < PIPE_SHADER_COMPUTE; s++)
+        for (int s = 0; s < MESA_SHADER_COMPUTE; s++)
                 v3d_predraw_check_stage_inputs(pctx, s);
 
         if (indirect && indirect->buffer) {
@@ -1197,7 +1197,7 @@ v3d_draw_vbo(struct pipe_context *pctx, const struct pipe_draw_info *info,
         /* Mark SSBOs and images as being written.  We don't actually know
          * which ones are read vs written, so just assume the worst.
          */
-        for (int s = 0; s < PIPE_SHADER_COMPUTE; s++) {
+        for (int s = 0; s < MESA_SHADER_COMPUTE; s++) {
                 unsigned i;
                 BITSET_FOREACH_SET(i, v3d->ssbo[s].enabled_mask,
                                    PIPE_MAX_SHADER_BUFFERS) {
@@ -1427,7 +1427,7 @@ v3d_launch_grid(struct pipe_context *pctx, const struct pipe_grid_info *info)
 
         MESA_TRACE_FUNC();
 
-        v3d_predraw_check_stage_inputs(pctx, PIPE_SHADER_COMPUTE);
+        v3d_predraw_check_stage_inputs(pctx, MESA_SHADER_COMPUTE);
 
         v3d_update_compiled_cs(v3d);
 
@@ -1559,7 +1559,7 @@ v3d_launch_grid(struct pipe_context *pctx, const struct pipe_grid_info *info)
 
         struct v3d_cl_reloc uniforms = v3d_write_uniforms(v3d, job,
                                                           v3d->prog.compute,
-                                                          PIPE_SHADER_COMPUTE);
+                                                          MESA_SHADER_COMPUTE);
         v3d_job_add_bo(job, uniforms.bo);
         submit.cfg[6] = uniforms.bo->offset + uniforms.offset;
 
@@ -1603,19 +1603,19 @@ v3d_launch_grid(struct pipe_context *pctx, const struct pipe_grid_info *info)
         /* Mark SSBOs as being written.. we don't actually know which ones are
          * read vs written, so just assume the worst
          */
-        BITSET_FOREACH_SET(i, v3d->ssbo[PIPE_SHADER_COMPUTE].enabled_mask,
+        BITSET_FOREACH_SET(i, v3d->ssbo[MESA_SHADER_COMPUTE].enabled_mask,
                            PIPE_MAX_SHADER_BUFFERS) {
                 struct v3d_resource *rsc = v3d_resource(
-                        v3d->ssbo[PIPE_SHADER_COMPUTE].sb[i].buffer);
+                        v3d->ssbo[MESA_SHADER_COMPUTE].sb[i].buffer);
                 rsc->writes++;
                 rsc->compute_written = true;
         }
 
         BITSET_FOREACH_SET(i,
-                           v3d->shaderimg[PIPE_SHADER_COMPUTE].enabled_mask,
+                           v3d->shaderimg[MESA_SHADER_COMPUTE].enabled_mask,
                            PIPE_MAX_SHADER_IMAGES) {
                 struct v3d_resource *rsc = v3d_resource(
-                        v3d->shaderimg[PIPE_SHADER_COMPUTE].si[i].base.resource);
+                        v3d->shaderimg[MESA_SHADER_COMPUTE].si[i].base.resource);
                 rsc->writes++;
                 rsc->compute_written = true;
         }
diff --git a/src/gallium/drivers/virgl/virgl_context.c b/src/gallium/drivers/virgl/virgl_context.c
index 41689d3ead8..38e0c6ac60b 100644
--- a/src/gallium/drivers/virgl/virgl_context.c
+++ b/src/gallium/drivers/virgl/virgl_context.c
@@ -321,7 +321,7 @@ static void virgl_reemit_draw_resources(struct virgl_context *vctx)
    /* framebuffer, sampler views, vertex/index/uniform/stream buffers */
    virgl_attach_res_framebuffer(vctx);
 
-   for (shader_type = 0; shader_type < PIPE_SHADER_COMPUTE; shader_type++) {
+   for (shader_type = 0; shader_type < MESA_SHADER_COMPUTE; shader_type++) {
       virgl_attach_res_sampler_views(vctx, shader_type);
       virgl_attach_res_uniform_buffers(vctx, shader_type);
       virgl_attach_res_shader_buffers(vctx, shader_type);
@@ -334,10 +334,10 @@ static void virgl_reemit_draw_resources(struct virgl_context *vctx)
 
 static void virgl_reemit_compute_resources(struct virgl_context *vctx)
 {
-   virgl_attach_res_sampler_views(vctx, PIPE_SHADER_COMPUTE);
-   virgl_attach_res_uniform_buffers(vctx, PIPE_SHADER_COMPUTE);
-   virgl_attach_res_shader_buffers(vctx, PIPE_SHADER_COMPUTE);
-   virgl_attach_res_shader_images(vctx, PIPE_SHADER_COMPUTE);
+   virgl_attach_res_sampler_views(vctx, MESA_SHADER_COMPUTE);
+   virgl_attach_res_uniform_buffers(vctx, MESA_SHADER_COMPUTE);
+   virgl_attach_res_shader_buffers(vctx, MESA_SHADER_COMPUTE);
+   virgl_attach_res_shader_images(vctx, MESA_SHADER_COMPUTE);
 
    virgl_attach_res_atomic_buffers(vctx);
 }
@@ -1410,7 +1410,7 @@ static void virgl_set_shader_buffers(struct pipe_context *ctx,
       }
    }
 
-   uint32_t max_shader_buffer = (shader == MESA_SHADER_FRAGMENT || shader == PIPE_SHADER_COMPUTE) ?
+   uint32_t max_shader_buffer = (shader == MESA_SHADER_FRAGMENT || shader == MESA_SHADER_COMPUTE) ?
       rs->caps.caps.v2.max_shader_buffer_frag_compute :
       rs->caps.caps.v2.max_shader_buffer_other_stages;
    if (!max_shader_buffer)
@@ -1469,7 +1469,7 @@ static void virgl_set_shader_images(struct pipe_context *ctx,
       }
    }
 
-   uint32_t max_shader_images = (shader == MESA_SHADER_FRAGMENT || shader == PIPE_SHADER_COMPUTE) ?
+   uint32_t max_shader_images = (shader == MESA_SHADER_FRAGMENT || shader == MESA_SHADER_COMPUTE) ?
      rs->caps.caps.v2.max_shader_image_frag_compute :
      rs->caps.caps.v2.max_shader_image_other_stages;
    if (!max_shader_images)
@@ -1519,7 +1519,7 @@ static void *virgl_create_compute_state(struct pipe_context *ctx,
       return NULL;
 
    handle = virgl_object_assign_handle();
-   ret = virgl_encode_shader_state(vctx, handle, PIPE_SHADER_COMPUTE,
+   ret = virgl_encode_shader_state(vctx, handle, MESA_SHADER_COMPUTE,
                                    &so_info,
                                    state->static_shared_mem,
                                    new_tokens);
@@ -1539,7 +1539,7 @@ static void virgl_bind_compute_state(struct pipe_context *ctx, void *state)
    uint32_t handle = (unsigned long)state;
    struct virgl_context *vctx = virgl_context(ctx);
 
-   virgl_encode_bind_shader(vctx, handle, PIPE_SHADER_COMPUTE);
+   virgl_encode_bind_shader(vctx, handle, MESA_SHADER_COMPUTE);
 }
 
 static void virgl_delete_compute_state(struct pipe_context *ctx, void *state)
diff --git a/src/gallium/drivers/virgl/virgl_encode.c b/src/gallium/drivers/virgl/virgl_encode.c
index 0dabe4469d5..487d4265882 100644
--- a/src/gallium/drivers/virgl/virgl_encode.c
+++ b/src/gallium/drivers/virgl/virgl_encode.c
@@ -821,7 +821,7 @@ int virgl_encode_shader_state(struct virgl_context *ctx,
 
       virgl_emit_shader_header(ctx, handle, len, virgl_shader_stage_convert(type), offlen, num_tokens);
 
-      if (type == PIPE_SHADER_COMPUTE)
+      if (type == MESA_SHADER_COMPUTE)
          virgl_encoder_write_dword(ctx->cbuf, cs_req_local_mem);
       else
          virgl_emit_shader_streamout(ctx, first_pass ? so_info : NULL);
@@ -1483,7 +1483,7 @@ int virgl_encode_link_shader(struct virgl_context *ctx, uint32_t *handles)
    virgl_encoder_write_dword(ctx->cbuf, handles[MESA_SHADER_GEOMETRY]);
    virgl_encoder_write_dword(ctx->cbuf, handles[MESA_SHADER_TESS_CTRL]);
    virgl_encoder_write_dword(ctx->cbuf, handles[MESA_SHADER_TESS_EVAL]);
-   virgl_encoder_write_dword(ctx->cbuf, handles[PIPE_SHADER_COMPUTE]);
+   virgl_encoder_write_dword(ctx->cbuf, handles[MESA_SHADER_COMPUTE]);
    return 0;
 }
 
diff --git a/src/gallium/drivers/virgl/virgl_screen.c b/src/gallium/drivers/virgl/virgl_screen.c
index 7252e2ba698..ef8dc097365 100644
--- a/src/gallium/drivers/virgl/virgl_screen.c
+++ b/src/gallium/drivers/virgl/virgl_screen.c
@@ -169,7 +169,7 @@ virgl_get_video_param(struct pipe_screen *screen,
 static void
 virgl_init_shader_caps(struct virgl_screen *vscreen)
 {
-   for (unsigned i = 0; i <= PIPE_SHADER_COMPUTE; i++) {
+   for (unsigned i = 0; i <= MESA_SHADER_COMPUTE; i++) {
       struct pipe_shader_caps *caps =
          (struct pipe_shader_caps *)&vscreen->base.shader_caps[i];
 
@@ -179,7 +179,7 @@ virgl_init_shader_caps(struct virgl_screen *vscreen)
          if (!vscreen->caps.caps.v1.bset.has_tessellation_shaders)
             continue;
          break;
-      case PIPE_SHADER_COMPUTE:
+      case MESA_SHADER_COMPUTE:
          if (!(vscreen->caps.caps.v2.capability_bits & VIRGL_CAP_COMPUTE_SHADER))
             continue;
          break;
@@ -238,14 +238,14 @@ virgl_init_shader_caps(struct virgl_screen *vscreen)
       int max_shader_buffers = VIRGL_SHADER_STAGE_CAP_V2(max_shader_storage_blocks, i);
       if (max_shader_buffers != INT_MAX) {
          caps->max_shader_buffers = max_shader_buffers;
-      } else if (i == MESA_SHADER_FRAGMENT || i == PIPE_SHADER_COMPUTE) {
+      } else if (i == MESA_SHADER_FRAGMENT || i == MESA_SHADER_COMPUTE) {
          caps->max_shader_buffers = vscreen->caps.caps.v2.max_shader_buffer_frag_compute;
       } else {
          caps->max_shader_buffers = vscreen->caps.caps.v2.max_shader_buffer_other_stages;
       }
 
       caps->max_shader_images =
-         i == MESA_SHADER_FRAGMENT || i == PIPE_SHADER_COMPUTE ?
+         i == MESA_SHADER_FRAGMENT || i == MESA_SHADER_COMPUTE ?
          vscreen->caps.caps.v2.max_shader_image_frag_compute :
          vscreen->caps.caps.v2.max_shader_image_other_stages;
 
diff --git a/src/gallium/drivers/virgl/virgl_screen.h b/src/gallium/drivers/virgl/virgl_screen.h
index 2f81c497cb2..fa36439dbcc 100644
--- a/src/gallium/drivers/virgl/virgl_screen.h
+++ b/src/gallium/drivers/virgl/virgl_screen.h
@@ -103,7 +103,7 @@ virgl_shader_stage_convert(enum pipe_shader_type type)
       return VIRGL_SHADER_GEOMETRY;
    case MESA_SHADER_FRAGMENT:
       return VIRGL_SHADER_FRAGMENT;
-   case PIPE_SHADER_COMPUTE:
+   case MESA_SHADER_COMPUTE:
       return VIRGL_SHADER_COMPUTE;
    default:
       UNREACHABLE("virgl: unknown shader stage.\n");
diff --git a/src/gallium/drivers/zink/zink_screen.c b/src/gallium/drivers/zink/zink_screen.c
index 7c185815636..b6692b09119 100644
--- a/src/gallium/drivers/zink/zink_screen.c
+++ b/src/gallium/drivers/zink/zink_screen.c
@@ -496,7 +496,7 @@ have_fp32_filter_linear(struct zink_screen *screen)
 static void
 zink_init_shader_caps(struct zink_screen *screen)
 {
-   for (unsigned i = 0; i <= PIPE_SHADER_COMPUTE; i++) {
+   for (unsigned i = 0; i <= MESA_SHADER_COMPUTE; i++) {
       struct pipe_shader_caps *caps =
          (struct pipe_shader_caps *)&screen->base.shader_caps[i];
 
diff --git a/src/gallium/frontends/rusticl/core/device.rs b/src/gallium/frontends/rusticl/core/device.rs
index 7cd8ac5a129..457162ba9c2 100644
--- a/src/gallium/frontends/rusticl/core/device.rs
+++ b/src/gallium/frontends/rusticl/core/device.rs
@@ -113,7 +113,7 @@ impl DeviceCaps {
     }
 
     fn shader_caps(screen: &PipeScreen) -> &pipe_shader_caps {
-        screen.shader_caps(pipe_shader_type::PIPE_SHADER_COMPUTE)
+        screen.shader_caps(pipe_shader_type::MESA_SHADER_COMPUTE)
     }
 }
 
@@ -777,7 +777,7 @@ impl DeviceBase {
 
     fn shader_caps(&self) -> &pipe_shader_caps {
         self.screen
-            .shader_caps(pipe_shader_type::PIPE_SHADER_COMPUTE)
+            .shader_caps(pipe_shader_type::MESA_SHADER_COMPUTE)
     }
 
     pub fn address_bits(&self) -> cl_uint {
@@ -892,7 +892,7 @@ impl DeviceBase {
         unsafe {
             *self
                 .screen
-                .nir_shader_compiler_options(pipe_shader_type::PIPE_SHADER_COMPUTE)
+                .nir_shader_compiler_options(pipe_shader_type::MESA_SHADER_COMPUTE)
         }
     }
 
@@ -1351,7 +1351,7 @@ impl Device {
     fn check_valid(screen: &PipeScreen) -> bool {
         if !screen.caps().compute
             || screen
-                .shader_caps(pipe_shader_type::PIPE_SHADER_COMPUTE)
+                .shader_caps(pipe_shader_type::MESA_SHADER_COMPUTE)
                 .supported_irs
                 & (1 << (pipe_shader_ir::PIPE_SHADER_IR_NIR as i32))
                 == 0
@@ -1362,7 +1362,7 @@ impl Device {
         // CL_DEVICE_MAX_PARAMETER_SIZE
         // For this minimum value, only a maximum of 128 arguments can be passed to a kernel
         if screen
-            .shader_caps(pipe_shader_type::PIPE_SHADER_COMPUTE)
+            .shader_caps(pipe_shader_type::MESA_SHADER_COMPUTE)
             .max_const_buffer0_size
             < 128
         {
diff --git a/src/gallium/frontends/rusticl/core/kernel.rs b/src/gallium/frontends/rusticl/core/kernel.rs
index add9334b249..16a998bf42c 100644
--- a/src/gallium/frontends/rusticl/core/kernel.rs
+++ b/src/gallium/frontends/rusticl/core/kernel.rs
@@ -553,7 +553,7 @@ impl CompilationResult {
         let nir = NirShader::deserialize(
             reader,
             d.screen()
-                .nir_shader_compiler_options(pipe_shader_type::PIPE_SHADER_COMPUTE),
+                .nir_shader_compiler_options(pipe_shader_type::MESA_SHADER_COMPUTE),
         )?;
         let compiled_args = CompiledKernelArg::deserialize(reader)?;
 
@@ -573,7 +573,7 @@ fn opt_nir(nir: &mut NirShader, dev: &Device, has_explicit_types: bool) {
     let nir_options = unsafe {
         &*dev
             .screen
-            .nir_shader_compiler_options(pipe_shader_type::PIPE_SHADER_COMPUTE)
+            .nir_shader_compiler_options(pipe_shader_type::MESA_SHADER_COMPUTE)
     };
 
     while {
@@ -813,7 +813,7 @@ fn compile_nir_variant(
     let nir_options = unsafe {
         &*dev
             .screen
-            .nir_shader_compiler_options(pipe_shader_type::PIPE_SHADER_COMPUTE)
+            .nir_shader_compiler_options(pipe_shader_type::MESA_SHADER_COMPUTE)
     };
 
     if variant == NirKernelVariant::Optimized {
diff --git a/src/gallium/frontends/rusticl/core/program.rs b/src/gallium/frontends/rusticl/core/program.rs
index 9ff363be053..29435f4bc0e 100644
--- a/src/gallium/frontends/rusticl/core/program.rs
+++ b/src/gallium/frontends/rusticl/core/program.rs
@@ -214,7 +214,7 @@ impl ProgramBuild {
         let nir = info.spirv.as_ref().unwrap().to_nir(
             kernel,
             d.screen
-                .nir_shader_compiler_options(pipe_shader_type::PIPE_SHADER_COMPUTE),
+                .nir_shader_compiler_options(pipe_shader_type::MESA_SHADER_COMPUTE),
             &d.spirv_caps,
             &d.lib_clc,
             &mut spec_constants,
diff --git a/src/gallium/frontends/rusticl/mesa/compiler/clc/spirv.rs b/src/gallium/frontends/rusticl/mesa/compiler/clc/spirv.rs
index 7a7059a4651..71aec1effa5 100644
--- a/src/gallium/frontends/rusticl/mesa/compiler/clc/spirv.rs
+++ b/src/gallium/frontends/rusticl/mesa/compiler/clc/spirv.rs
@@ -350,7 +350,7 @@ impl SPIRVBin {
     }
 
     pub fn get_lib_clc(screen: &PipeScreen, spirv_caps: &spirv_capabilities) -> Option<NirShader> {
-        let nir_options = screen.nir_shader_compiler_options(pipe_shader_type::PIPE_SHADER_COMPUTE);
+        let nir_options = screen.nir_shader_compiler_options(pipe_shader_type::MESA_SHADER_COMPUTE);
         let address_bits = screen.compute_caps().address_bits;
         let spirv_options =
             Self::get_spirv_options(false, ptr::null(), address_bits, spirv_caps, None);
diff --git a/src/gallium/frontends/rusticl/mesa/pipe/context.rs b/src/gallium/frontends/rusticl/mesa/pipe/context.rs
index 149988f6a28..57b4726644b 100644
--- a/src/gallium/frontends/rusticl/mesa/pipe/context.rs
+++ b/src/gallium/frontends/rusticl/mesa/pipe/context.rs
@@ -394,7 +394,7 @@ impl PipeContext {
         unsafe {
             self.pipe.as_ref().bind_sampler_states.unwrap()(
                 self.pipe.as_ptr(),
-                pipe_shader_type::PIPE_SHADER_COMPUTE,
+                pipe_shader_type::MESA_SHADER_COMPUTE,
                 0,
                 samplers.len() as u32,
                 samplers.as_mut_ptr(),
@@ -407,7 +407,7 @@ impl PipeContext {
         unsafe {
             self.pipe.as_ref().bind_sampler_states.unwrap()(
                 self.pipe.as_ptr(),
-                pipe_shader_type::PIPE_SHADER_COMPUTE,
+                pipe_shader_type::MESA_SHADER_COMPUTE,
                 0,
                 count,
                 samplers.as_mut_ptr(),
@@ -429,7 +429,7 @@ impl PipeContext {
         unsafe {
             self.pipe.as_ref().set_constant_buffer.unwrap()(
                 self.pipe.as_ptr(),
-                pipe_shader_type::PIPE_SHADER_COMPUTE,
+                pipe_shader_type::MESA_SHADER_COMPUTE,
                 idx,
                 false,
                 &cb,
@@ -447,7 +447,7 @@ impl PipeContext {
         unsafe {
             self.pipe.as_ref().set_constant_buffer.unwrap()(
                 self.pipe.as_ptr(),
-                pipe_shader_type::PIPE_SHADER_COMPUTE,
+                pipe_shader_type::MESA_SHADER_COMPUTE,
                 idx,
                 false,
                 if data.is_empty() { ptr::null() } else { &cb },
@@ -484,7 +484,7 @@ impl PipeContext {
 
             self.pipe.as_ref().set_constant_buffer.unwrap()(
                 self.pipe.as_ptr(),
-                pipe_shader_type::PIPE_SHADER_COMPUTE,
+                pipe_shader_type::MESA_SHADER_COMPUTE,
                 idx,
                 true,
                 &cb,
@@ -544,7 +544,7 @@ impl PipeContext {
         unsafe {
             self.pipe.as_ref().set_sampler_views.unwrap()(
                 self.pipe.as_ptr(),
-                pipe_shader_type::PIPE_SHADER_COMPUTE,
+                pipe_shader_type::MESA_SHADER_COMPUTE,
                 0,
                 views.len() as u32,
                 unbind_trailing,
@@ -558,7 +558,7 @@ impl PipeContext {
         unsafe {
             self.pipe.as_ref().set_sampler_views.unwrap()(
                 self.pipe.as_ptr(),
-                pipe_shader_type::PIPE_SHADER_COMPUTE,
+                pipe_shader_type::MESA_SHADER_COMPUTE,
                 0,
                 count,
                 0,
@@ -572,7 +572,7 @@ impl PipeContext {
         unsafe {
             self.pipe.as_ref().set_shader_images.unwrap()(
                 self.pipe.as_ptr(),
-                pipe_shader_type::PIPE_SHADER_COMPUTE,
+                pipe_shader_type::MESA_SHADER_COMPUTE,
                 0,
                 images.len() as u32,
                 unbind_trailing,
@@ -585,7 +585,7 @@ impl PipeContext {
         unsafe {
             self.pipe.as_ref().set_shader_images.unwrap()(
                 self.pipe.as_ptr(),
-                pipe_shader_type::PIPE_SHADER_COMPUTE,
+                pipe_shader_type::MESA_SHADER_COMPUTE,
                 0,
                 count,
                 0,
diff --git a/src/mesa/state_tracker/st_atom_constbuf.c b/src/mesa/state_tracker/st_atom_constbuf.c
index 7ba9151d9df..880d4c2da9b 100644
--- a/src/mesa/state_tracker/st_atom_constbuf.c
+++ b/src/mesa/state_tracker/st_atom_constbuf.c
@@ -82,7 +82,7 @@ st_upload_constants(struct st_context *st, struct gl_program *prog, gl_shader_st
           shader_type == MESA_SHADER_GEOMETRY ||
           shader_type == MESA_SHADER_TESS_CTRL ||
           shader_type == MESA_SHADER_TESS_EVAL ||
-          shader_type == PIPE_SHADER_COMPUTE);
+          shader_type == MESA_SHADER_COMPUTE);
 
    /* update the ATI constants before rendering */
    if (shader_type == MESA_SHADER_FRAGMENT && prog->ati_fs) {
@@ -359,5 +359,5 @@ st_bind_cs_ubos(struct st_context *st)
    struct gl_program *prog =
       st->ctx->_Shader->CurrentProgram[MESA_SHADER_COMPUTE];
 
-   st_bind_ubos(st, prog, PIPE_SHADER_COMPUTE);
+   st_bind_ubos(st, prog, MESA_SHADER_COMPUTE);
 }
diff --git a/src/mesa/state_tracker/st_atom_image.c b/src/mesa/state_tracker/st_atom_image.c
index 7a60d1b3b55..cde0e3cf4a4 100644
--- a/src/mesa/state_tracker/st_atom_image.c
+++ b/src/mesa/state_tracker/st_atom_image.c
@@ -220,5 +220,5 @@ void st_bind_cs_images(struct st_context *st)
    struct gl_program *prog =
       st->ctx->_Shader->CurrentProgram[MESA_SHADER_COMPUTE];
 
-   st_bind_images(st, prog, PIPE_SHADER_COMPUTE);
+   st_bind_images(st, prog, MESA_SHADER_COMPUTE);
 }
diff --git a/src/mesa/state_tracker/st_atom_sampler.c b/src/mesa/state_tracker/st_atom_sampler.c
index 6055f5c9bbd..ada28200a32 100644
--- a/src/mesa/state_tracker/st_atom_sampler.c
+++ b/src/mesa/state_tracker/st_atom_sampler.c
@@ -424,7 +424,7 @@ st_update_compute_samplers(struct st_context *st)
 
    if (ctx->ComputeProgram._Current) {
       update_shader_samplers(st,
-                             PIPE_SHADER_COMPUTE,
+                             MESA_SHADER_COMPUTE,
                              ctx->ComputeProgram._Current, NULL, NULL);
    }
 }
diff --git a/src/mesa/state_tracker/st_atom_shader.c b/src/mesa/state_tracker/st_atom_shader.c
index 79bac394fad..e899572c888 100644
--- a/src/mesa/state_tracker/st_atom_shader.c
+++ b/src/mesa/state_tracker/st_atom_shader.c
@@ -344,6 +344,6 @@ st_update_cp(struct st_context *st)
 {
    void *shader = st_update_common_program(st,
                                            st->ctx->ComputeProgram._Current,
-                                           PIPE_SHADER_COMPUTE, &st->cp);
+                                           MESA_SHADER_COMPUTE, &st->cp);
    cso_set_compute_shader_handle(st->cso_context, shader);
 }
diff --git a/src/mesa/state_tracker/st_atom_storagebuf.c b/src/mesa/state_tracker/st_atom_storagebuf.c
index eb13b101326..f4e960ac563 100644
--- a/src/mesa/state_tracker/st_atom_storagebuf.c
+++ b/src/mesa/state_tracker/st_atom_storagebuf.c
@@ -136,5 +136,5 @@ void st_bind_cs_ssbos(struct st_context *st)
    struct gl_program *prog =
       st->ctx->_Shader->CurrentProgram[MESA_SHADER_COMPUTE];
 
-   st_bind_ssbos(st, prog, PIPE_SHADER_COMPUTE);
+   st_bind_ssbos(st, prog, MESA_SHADER_COMPUTE);
 }
diff --git a/src/mesa/state_tracker/st_atom_texture.c b/src/mesa/state_tracker/st_atom_texture.c
index 7cd7626ed81..ff07bb44068 100644
--- a/src/mesa/state_tracker/st_atom_texture.c
+++ b/src/mesa/state_tracker/st_atom_texture.c
@@ -401,7 +401,7 @@ st_update_compute_textures(struct st_context *st)
    const struct gl_context *ctx = st->ctx;
 
    if (ctx->ComputeProgram._Current) {
-      update_textures(st, PIPE_SHADER_COMPUTE,
+      update_textures(st, MESA_SHADER_COMPUTE,
                             ctx->ComputeProgram._Current);
    }
 }
diff --git a/src/mesa/state_tracker/st_context.c b/src/mesa/state_tracker/st_context.c
index 288b8b613d0..0d398d1af37 100644
--- a/src/mesa/state_tracker/st_context.c
+++ b/src/mesa/state_tracker/st_context.c
@@ -302,7 +302,7 @@ free_zombie_shaders(struct st_context *st)
          st->ctx->NewDriverState |= ST_NEW_TES_STATE;
          st->pipe->delete_tes_state(st->pipe, entry->shader);
          break;
-      case PIPE_SHADER_COMPUTE:
+      case MESA_SHADER_COMPUTE:
          st->ctx->NewDriverState |= ST_NEW_CS_STATE;
          st->pipe->delete_compute_state(st->pipe, entry->shader);
          break;
diff --git a/src/mesa/state_tracker/st_extensions.c b/src/mesa/state_tracker/st_extensions.c
index f859490a7cc..b0cfacd7b4a 100644
--- a/src/mesa/state_tracker/st_extensions.c
+++ b/src/mesa/state_tracker/st_extensions.c
@@ -206,7 +206,7 @@ void st_init_limits(struct pipe_screen *screen,
       struct gl_program_constants *pc = &c->Program[stage];
 
       if (!screen->nir_options[stage] ||
-          (sh == PIPE_SHADER_COMPUTE && !screen->caps.compute))
+          (sh == MESA_SHADER_COMPUTE && !screen->caps.compute))
          continue;
 
       pc->MaxTextureImageUnits =
diff --git a/src/mesa/state_tracker/st_pbo_compute.c b/src/mesa/state_tracker/st_pbo_compute.c
index f4f4b269964..95d19ca333f 100644
--- a/src/mesa/state_tracker/st_pbo_compute.c
+++ b/src/mesa/state_tracker/st_pbo_compute.c
@@ -1004,7 +1004,7 @@ download_texture_compute(struct st_context *st,
    assert(cs);
    struct cso_context *cso = st->cso_context;
 
-   pipe->set_constant_buffer(pipe, PIPE_SHADER_COMPUTE, 0, false, &cb);
+   pipe->set_constant_buffer(pipe, MESA_SHADER_COMPUTE, 0, false, &cb);
 
    cso_save_compute_state(cso, CSO_BIT_COMPUTE_SHADER | CSO_BIT_COMPUTE_SAMPLERS);
    cso_set_compute_shader_handle(cso, cs);
@@ -1112,12 +1112,12 @@ download_texture_compute(struct st_context *st,
       if (sampler_view == NULL)
          goto fail;
 
-      pipe->set_sampler_views(pipe, PIPE_SHADER_COMPUTE, 0, 1, 0,
+      pipe->set_sampler_views(pipe, MESA_SHADER_COMPUTE, 0, 1, 0,
                               &sampler_view);
-      st->state.num_sampler_views[PIPE_SHADER_COMPUTE] =
-         MAX2(st->state.num_sampler_views[PIPE_SHADER_COMPUTE], 1);
+      st->state.num_sampler_views[MESA_SHADER_COMPUTE] =
+         MAX2(st->state.num_sampler_views[MESA_SHADER_COMPUTE], 1);
 
-      cso_set_samplers(cso, PIPE_SHADER_COMPUTE, 1, samplers);
+      cso_set_samplers(cso, MESA_SHADER_COMPUTE, 1, samplers);
    }
 
    /* Set up destination buffer */
@@ -1143,7 +1143,7 @@ download_texture_compute(struct st_context *st,
       buffer.buffer = dst;
       buffer.buffer_size = buffer_size;
 
-      pipe->set_shader_buffers(pipe, PIPE_SHADER_COMPUTE, 0, 1, &buffer, 0x1);
+      pipe->set_shader_buffers(pipe, MESA_SHADER_COMPUTE, 0, 1, &buffer, 0x1);
    }
 
    struct pipe_grid_info info = { 0 };
@@ -1165,11 +1165,11 @@ fail:
    /* Unbind all because st/mesa won't do it if the current shader doesn't
     * use them.
     */
-   pipe->set_sampler_views(pipe, PIPE_SHADER_COMPUTE, 0, 0,
-                           st->state.num_sampler_views[PIPE_SHADER_COMPUTE],
+   pipe->set_sampler_views(pipe, MESA_SHADER_COMPUTE, 0, 0,
+                           st->state.num_sampler_views[MESA_SHADER_COMPUTE],
                            NULL);
-   st->state.num_sampler_views[PIPE_SHADER_COMPUTE] = 0;
-   pipe->set_shader_buffers(pipe, PIPE_SHADER_COMPUTE, 0, 1, NULL, 0);
+   st->state.num_sampler_views[MESA_SHADER_COMPUTE] = 0;
+   pipe->set_shader_buffers(pipe, MESA_SHADER_COMPUTE, 0, 1, NULL, 0);
 
    st->ctx->NewDriverState |= ST_NEW_CS_CONSTANTS |
                               ST_NEW_CS_SSBOS |
diff --git a/src/mesa/state_tracker/st_texcompress_compute.c b/src/mesa/state_tracker/st_texcompress_compute.c
index 3f4fafc5d45..b3da46b5750 100644
--- a/src/mesa/state_tracker/st_texcompress_compute.c
+++ b/src/mesa/state_tracker/st_texcompress_compute.c
@@ -150,12 +150,12 @@ bind_compute_state(struct st_context *st,
                    bool cs_handle_from_prog,
                    bool constbuf0_from_prog)
 {
-   assert(prog->info.stage == PIPE_SHADER_COMPUTE);
+   assert(prog->info.stage == MESA_SHADER_COMPUTE);
 
    /* Set compute states in the same order as defined in st_atom_list.h */
 
    assert(prog->affected_states & ST_NEW_CS_STATE);
-   assert(st->shader_has_one_variant[PIPE_SHADER_COMPUTE]);
+   assert(st->shader_has_one_variant[MESA_SHADER_COMPUTE]);
    cso_set_compute_shader_handle(st->cso_context,
                                  cs_handle_from_prog ?
                                  prog->variants->driver_shader : NULL);
@@ -217,7 +217,7 @@ dispatch_compute_state(struct st_context *st,
                        unsigned num_workgroups_y,
                        unsigned num_workgroups_z)
 {
-   assert(prog->info.stage == PIPE_SHADER_COMPUTE);
+   assert(prog->info.stage == MESA_SHADER_COMPUTE);
 
    /* Bind the state */
    bind_compute_state(st, prog, sampler_views, shader_buffers, image_views,