fdo-mirrors/mesa
1
0
Fork 0
mirror of https://gitlab.freedesktop.org/mesa/mesa.git synced 2026-04-20 23:00:36 +02:00
Code Issues Projects Releases Packages Wiki Activity Actions 1

aco: disable shared VGPRs for non-monolithic shaders on GFX9+

Browse source 
For unmerged shaders on GFX9+, we would need to jump to the second
shader part which means shared VGPRs can't be enabled.

Signed-off-by: Samuel Pitoiset <samuel.pitoiset@gmail.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/24697>
This commit is contained in:
Samuel Pitoiset 2023-08-15 10:52:53 +02:00 • committed by Marge Bot
parent 4224da6726
commit aba16211a8
3 changed files with 7 additions and 4 deletions
Show stats Download patch file Download diff file Expand all files Collapse all files

9
src/amd/compiler/aco_instruction_selection.cpp
View file

@ -196,10 +196,11 @@ emit_bpermute(isel_context* ctx, Builder& bld, Temp index, Temp data)
* of multiple binaries, because the VGPR use is not known when choosing
* which registers to use for the shared VGPRs.
*/
const bool avoid_shared_vgprs = ctx->options->gfx_level >= GFX10 &&
ctx->options->gfx_level < GFX11 &&
ctx->program->wave_size == 64 &&
(ctx->program->info.has_epilog || ctx->stage == raytracing_cs);
const bool avoid_shared_vgprs =
ctx->options->gfx_level >= GFX10 && ctx->options->gfx_level < GFX11 &&
ctx->program->wave_size == 64 &&
(ctx->program->info.has_epilog || !ctx->program->info.is_monolithic ||
ctx->stage == raytracing_cs);
if (ctx->options->gfx_level <= GFX7 || avoid_shared_vgprs) {
/* GFX6-7: there is no bpermute instruction */

1
src/amd/compiler/aco_shader_info.h
View file

@ -102,6 +102,7 @@ struct aco_shader_info {
bool image_2d_view_of_3d;
unsigned workgroup_size;
bool has_epilog; /* Only for TCS or PS. */
bool is_monolithic;
struct {
bool tcs_in_out_eq;
uint64_t tcs_temp_only_input_mask;

1
src/amd/vulkan/radv_aco_shader_info.h
View file

@ -49,6 +49,7 @@ radv_aco_convert_shader_info(struct aco_shader_info *aco_info, const struct radv
ASSIGN_FIELD(has_ngg_early_prim_export);
ASSIGN_FIELD(workgroup_size);
ASSIGN_FIELD(has_epilog);
ASSIGN_FIELD(is_monolithic);
ASSIGN_FIELD(vs.tcs_in_out_eq);
ASSIGN_FIELD(vs.tcs_temp_only_input_mask);
ASSIGN_FIELD(vs.has_prolog);