radeonsi: stop using TGSI_PROPERTY_CS_*

Reviewed-by: Pierre-Eric Pelloux-Prayer <pierre-eric.pelloux-prayer@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/6624>

diff --git a/src/gallium/drivers/radeonsi/si_compute.c b/src/gallium/drivers/radeonsi/si_compute.c
index 4292b9d573736534045e6e7048a07b26f51a8e9d..4fc81ec464cf3d00b7e48a3c0b5262739c21b0ca 100644 (file)
--- a/src/gallium/drivers/radeonsi/si_compute.c
+++ b/src/gallium/drivers/radeonsi/si_compute.c
@@ -137,9 +137,9 @@ static void si_create_compute_state_async(void *job, int thread_index)
 
    program->shader.is_monolithic = true;
    program->reads_variable_block_size =
 
    program->shader.is_monolithic = true;
    program->reads_variable_block_size =

-      sel->info.uses_block_size && sel->info.properties[TGSI_PROPERTY_CS_FIXED_BLOCK_WIDTH] == 0;
+      sel->info.uses_block_size && sel->info.base.cs.local_size[0] == 0;

    program->num_cs_user_data_dwords =
    program->num_cs_user_data_dwords =

-      sel->info.properties[TGSI_PROPERTY_CS_USER_DATA_COMPONENTS_AMD];
+      sel->info.base.cs.user_data_components_amd;

 
    unsigned user_sgprs = SI_NUM_RESOURCE_SGPRS + (sel->info.uses_grid_size ? 3 : 0) +
                          (program->reads_variable_block_size ? 3 : 0) +
 
    unsigned user_sgprs = SI_NUM_RESOURCE_SGPRS + (sel->info.uses_grid_size ? 3 : 0) +
                          (program->reads_variable_block_size ? 3 : 0) +

diff --git a/src/gallium/drivers/radeonsi/si_shader.c b/src/gallium/drivers/radeonsi/si_shader.c
index c77585256573db6448ae06571e333963006eeb2b..f10195febd47c7eabb5c4ee9a4810d2a3a8fea2c 100644 (file)
--- a/src/gallium/drivers/radeonsi/si_shader.c
+++ b/src/gallium/drivers/radeonsi/si_shader.c
@@ -211,10 +211,10 @@ unsigned si_get_max_workgroup_size(const struct si_shader *shader)
       return 0;
    }
 
       return 0;
    }
 

-   const unsigned *properties = shader->selector->info.properties;
-   unsigned max_work_group_size = properties[TGSI_PROPERTY_CS_FIXED_BLOCK_WIDTH] *
-                                  properties[TGSI_PROPERTY_CS_FIXED_BLOCK_HEIGHT] *
-                                  properties[TGSI_PROPERTY_CS_FIXED_BLOCK_DEPTH];
+   uint16_t *local_size = shader->selector->info.base.cs.local_size;
+   unsigned max_work_group_size = (uint32_t)local_size[0] *
+                                  (uint32_t)local_size[1] *
+                                  (uint32_t)local_size[2];

 
    if (!max_work_group_size) {
       /* This is a variable group size compute shader,
 
    if (!max_work_group_size) {
       /* This is a variable group size compute shader,


@@ -696,11 +696,11 @@ void si_create_function(struct si_shader_context *ctx, bool ngg_cull_shader)
       if (shader->selector->info.uses_grid_size)
          ac_add_arg(&ctx->args, AC_ARG_SGPR, 3, AC_ARG_INT, &ctx->args.num_work_groups);
       if (shader->selector->info.uses_block_size &&
       if (shader->selector->info.uses_grid_size)
          ac_add_arg(&ctx->args, AC_ARG_SGPR, 3, AC_ARG_INT, &ctx->args.num_work_groups);
       if (shader->selector->info.uses_block_size &&

-          shader->selector->info.properties[TGSI_PROPERTY_CS_FIXED_BLOCK_WIDTH] == 0)
+          shader->selector->info.base.cs.local_size[0] == 0)

          ac_add_arg(&ctx->args, AC_ARG_SGPR, 3, AC_ARG_INT, &ctx->block_size);
 
       unsigned cs_user_data_dwords =
          ac_add_arg(&ctx->args, AC_ARG_SGPR, 3, AC_ARG_INT, &ctx->block_size);
 
       unsigned cs_user_data_dwords =

-         shader->selector->info.properties[TGSI_PROPERTY_CS_USER_DATA_COMPONENTS_AMD];
+         shader->selector->info.base.cs.user_data_components_amd;

       if (cs_user_data_dwords) {
          ac_add_arg(&ctx->args, AC_ARG_SGPR, cs_user_data_dwords, AC_ARG_INT, &ctx->cs_user_data);
       }
       if (cs_user_data_dwords) {
          ac_add_arg(&ctx->args, AC_ARG_SGPR, cs_user_data_dwords, AC_ARG_INT, &ctx->cs_user_data);
       }

diff --git a/src/gallium/drivers/radeonsi/si_shader_llvm.c b/src/gallium/drivers/radeonsi/si_shader_llvm.c
index 26aabf3e7fc2d0e2b338222b6e7af24427fedffe..27b5593cb5fa156c6ed4e9707e70ee2e1039759a 100644 (file)
--- a/src/gallium/drivers/radeonsi/si_shader_llvm.c
+++ b/src/gallium/drivers/radeonsi/si_shader_llvm.c
@@ -370,15 +370,11 @@ LLVMValueRef si_llvm_get_block_size(struct ac_shader_abi *abi)
    LLVMValueRef values[3];
    LLVMValueRef result;
    unsigned i;
    LLVMValueRef values[3];
    LLVMValueRef result;
    unsigned i;

-   unsigned *properties = ctx->shader->selector->info.properties;
-
-   if (properties[TGSI_PROPERTY_CS_FIXED_BLOCK_WIDTH] != 0) {
-      unsigned sizes[3] = {properties[TGSI_PROPERTY_CS_FIXED_BLOCK_WIDTH],
-                           properties[TGSI_PROPERTY_CS_FIXED_BLOCK_HEIGHT],
-                           properties[TGSI_PROPERTY_CS_FIXED_BLOCK_DEPTH]};
+   uint16_t *local_size = ctx->shader->selector->info.base.cs.local_size;

 
 

+   if (local_size[0] != 0) {

       for (i = 0; i < 3; ++i)
       for (i = 0; i < 3; ++i)

-         values[i] = LLVMConstInt(ctx->ac.i32, sizes[i], 0);
+         values[i] = LLVMConstInt(ctx->ac.i32, local_size[i], 0);

 
       result = ac_build_gather_values(&ctx->ac, values, 3);
    } else {
 
       result = ac_build_gather_values(&ctx->ac, values, 3);
    } else {

diff --git a/src/gallium/drivers/radeonsi/si_shader_nir.c b/src/gallium/drivers/radeonsi/si_shader_nir.c
index 3bbba5eacf4b7520a35ac5aaf4341b79a35825ec..ae859430291b7837e9dc5b8afc7b484a340ceec8 100644 (file)
--- a/src/gallium/drivers/radeonsi/si_shader_nir.c
+++ b/src/gallium/drivers/radeonsi/si_shader_nir.c
@@ -289,7 +289,7 @@ static void scan_instruction(const struct nir_shader *nir, struct si_shader_info
          break;
       case nir_intrinsic_load_local_group_size:
          /* The block size is translated to IMM with a fixed block size. */
          break;
       case nir_intrinsic_load_local_group_size:
          /* The block size is translated to IMM with a fixed block size. */

-         if (info->properties[TGSI_PROPERTY_CS_FIXED_BLOCK_WIDTH] == 0)
+         if (info->base.cs.local_size[0] == 0)

             info->uses_block_size = true;
          break;
       case nir_intrinsic_load_local_invocation_id:
             info->uses_block_size = true;
          break;
       case nir_intrinsic_load_local_invocation_id:


@@ -504,14 +504,6 @@ void si_nir_scan_shader(const struct nir_shader *nir, struct si_shader_info *inf
                                                                       TGSI_INTERPOLATE_LOC_CENTER;
    }
 
                                                                       TGSI_INTERPOLATE_LOC_CENTER;
    }
 

-   if (gl_shader_stage_is_compute(nir->info.stage)) {
-      info->properties[TGSI_PROPERTY_CS_FIXED_BLOCK_WIDTH] = nir->info.cs.local_size[0];
-      info->properties[TGSI_PROPERTY_CS_FIXED_BLOCK_HEIGHT] = nir->info.cs.local_size[1];
-      info->properties[TGSI_PROPERTY_CS_FIXED_BLOCK_DEPTH] = nir->info.cs.local_size[2];
-      info->properties[TGSI_PROPERTY_CS_USER_DATA_COMPONENTS_AMD] =
-         nir->info.cs.user_data_components_amd;
-   }
-

    info->constbuf0_num_slots = nir->num_uniforms;
    info->shader_buffers_declared = u_bit_consecutive(0, nir->info.num_ssbos);
    info->const_buffers_declared = u_bit_consecutive(0, nir->info.num_ubos);
    info->constbuf0_num_slots = nir->num_uniforms;
    info->shader_buffers_declared = u_bit_consecutive(0, nir->info.num_ssbos);
    info->const_buffers_declared = u_bit_consecutive(0, nir->info.num_ubos);