From 112e66fa090929401b2193e32a905221c3233a49 Mon Sep 17 00:00:00 2001 From: Samuel Pitoiset Date: Thu, 8 Oct 2020 13:51:27 +0200 Subject: [PATCH] aco: compute the CS workgroup size from the shader NIR info cs.block_size is copied from cs.local_size during the shader info pass. Signed-off-by: Samuel Pitoiset Reviewed-by: Rhys Perry Part-of: --- src/amd/compiler/aco_instruction_selection_setup.cpp | 5 +++-- 1 file changed, 3 insertions(+), 2 deletions(-) diff --git a/src/amd/compiler/aco_instruction_selection_setup.cpp b/src/amd/compiler/aco_instruction_selection_setup.cpp index b689f1d..e5b7092 100644 --- a/src/amd/compiler/aco_instruction_selection_setup.cpp +++ b/src/amd/compiler/aco_instruction_selection_setup.cpp @@ -1240,8 +1240,9 @@ setup_isel_context(Program* program, program->workgroup_size = program->wave_size; } else if (program->stage == compute_cs) { /* CS sets the workgroup size explicitly */ - unsigned* bsize = program->info->cs.block_size; - program->workgroup_size = bsize[0] * bsize[1] * bsize[2]; + program->workgroup_size = shaders[0]->info.cs.local_size[0] * + shaders[0]->info.cs.local_size[1] * + shaders[0]->info.cs.local_size[2]; } else if ((program->stage & hw_es) || program->stage == geometry_gs) { /* Unmerged ESGS operate in workgroups if on-chip GS (LDS rings) are enabled on GFX7-8 (not implemented in Mesa) */ program->workgroup_size = program->wave_size; -- 2.7.4