AlexIndustrial/mesa
1
0
Fork 0
Code Issues Pull Requests Actions Packages Projects Releases Wiki Activity

radeonsi: set threadgroup size to 0 for threadgroups with only 1 wave

Browse Source 
This has no effect on Wave64.

Reviewed-by: Pierre-Eric Pelloux-Prayer <pierre-eric.pelloux-prayer@amd.com>
Acked-by: Samuel Pitoiset <samuel.pitoiset@gmail.com>
This commit is contained in:
Marek Olšák
2019-07-12 17:22:30 -04:00
parent a8a526c5cb
commit e2c8ff009e
1 changed files with 3 additions and 3 deletions
Download Patch File Download Diff File Expand all files Collapse all files
src/gallium/drivers/radeonsi/si_shader.c
+3 -3
View File
@@ -4484,10 +4484,10 @@ static unsigned si_get_max_workgroup_size(const struct si_shader *shader)
case PIPE_SHADER_TESS_CTRL:
/* Return this so that LLVM doesn't remove s_barrier
* instructions on chips where we use s_barrier. */
return shader->selector->screen->info.chip_class >= GFX7 ? 128 : 64;
return shader->selector->screen->info.chip_class >= GFX7 ? 128 : 0;
case PIPE_SHADER_GEOMETRY:
return shader->selector->screen->info.chip_class >= GFX9 ? 128 : 64;
return shader->selector->screen->info.chip_class >= GFX9 ? 128 : 0;
case PIPE_SHADER_COMPUTE:
break; /* see below */
@@ -7626,7 +7626,7 @@ static void si_build_tcs_epilog_function(struct si_shader_context *ctx,
/* Create the function. */
si_create_function(ctx, "tcs_epilog", NULL, 0, &fninfo,
ctx->screen->info.chip_class >= GFX7 ? 128 : 64);
ctx->screen->info.chip_class >= GFX7 ? 128 : 0);
ac_declare_lds_as_pointer(&ctx->ac);
func = ctx->main_fn;