From c63ac28014695aac21bb1b78864a6d75ebbf215d Mon Sep 17 00:00:00 2001 From: Rhys Perry Date: Thu, 7 Sep 2023 17:01:14 +0100 Subject: [PATCH] ac/nir: optimize mesh shader local_invocation_index MIME-Version: 1.0 Content-Type: text/plain; charset=UTF-8 Content-Transfer-Encoding: 8bit Signed-off-by: Rhys Perry Reviewed-by: Timur Kristóf Part-of: --- src/amd/common/ac_nir_lower_ngg.c | 13 +++++++++++++ 1 file changed, 13 insertions(+) diff --git a/src/amd/common/ac_nir_lower_ngg.c b/src/amd/common/ac_nir_lower_ngg.c index 2257415c845..23911413f4f 100644 --- a/src/amd/common/ac_nir_lower_ngg.c +++ b/src/amd/common/ac_nir_lower_ngg.c @@ -4958,5 +4958,18 @@ ac_nir_lower_ngg_ms(nir_shader *shader, nir_lower_alu_to_scalar(shader, NULL, NULL); nir_lower_phis_to_scalar(shader, true); + /* Optimize load_local_invocation_index. When the API workgroup is smaller than the HW workgroup, + * local_invocation_id isn't initialized for all lanes and we can't perform this optimization for + * all load_local_invocation_index. + */ + if (fast_launch_2 && api_workgroup_size == hw_workgroup_size && + ((shader->info.workgroup_size[0] == 1) + (shader->info.workgroup_size[1] == 1) + + (shader->info.workgroup_size[2] == 1)) == 2) { + nir_lower_compute_system_values_options csv_options = { + .lower_local_invocation_index = true, + }; + nir_lower_compute_system_values(shader, &csv_options); + } + nir_validate_shader(shader, "after emitting NGG MS"); }