Module: Mesa Branch: master Commit: 112e66fa090929401b2193e32a905221c3233a49 URL: http://cgit.freedesktop.org/mesa/mesa/commit/?id=112e66fa090929401b2193e32a905221c3233a49
Author: Samuel Pitoiset <[email protected]> Date: Thu Oct 8 13:51:27 2020 +0200 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 <[email protected]> Reviewed-by: Rhys Perry <[email protected]> Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/7061> --- 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 b689f1d27cb..e5b70920307 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; _______________________________________________ mesa-commit mailing list [email protected] https://lists.freedesktop.org/mailman/listinfo/mesa-commit
