Module: Mesa Branch: main Commit: 6766e6a985d7a1ea3dc0eb3896c9fce745b9717b URL: http://cgit.freedesktop.org/mesa/mesa/commit/?id=6766e6a985d7a1ea3dc0eb3896c9fce745b9717b
Author: Timur Kristóf <[email protected]> Date: Thu Oct 21 11:33:10 2021 +0200 aco: Add Mesh and Task shader stages. Signed-off-by: Timur Kristóf <[email protected]> Reviewed-by: Rhys Perry <[email protected]> Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/13580> --- src/amd/compiler/aco_instruction_selection.cpp | 2 +- .../compiler/aco_instruction_selection_setup.cpp | 27 +++++++++++++++++++++- src/amd/compiler/aco_ir.h | 15 ++++++++---- src/amd/compiler/aco_print_ir.cpp | 4 ++++ 4 files changed, 41 insertions(+), 7 deletions(-) diff --git a/src/amd/compiler/aco_instruction_selection.cpp b/src/amd/compiler/aco_instruction_selection.cpp index 649bf1923fc..c4637a6d39e 100644 --- a/src/amd/compiler/aco_instruction_selection.cpp +++ b/src/amd/compiler/aco_instruction_selection.cpp @@ -4761,7 +4761,7 @@ void visit_store_output(isel_context* ctx, nir_intrinsic_instr* instr) { if (ctx->stage == vertex_vs || ctx->stage == tess_eval_vs || ctx->stage == fragment_fs || - ctx->stage == vertex_ngg || ctx->stage == tess_eval_ngg || + ctx->stage == vertex_ngg || ctx->stage == tess_eval_ngg || ctx->stage == mesh_ngg || (ctx->stage == vertex_tess_control_hs && ctx->shader->info.stage == MESA_SHADER_VERTEX) || ctx->shader->info.stage == MESA_SHADER_GEOMETRY) { bool stored_to_temps = store_output_to_temps(ctx, instr); diff --git a/src/amd/compiler/aco_instruction_selection_setup.cpp b/src/amd/compiler/aco_instruction_selection_setup.cpp index ed72d3009ba..f6c92482b1f 100644 --- a/src/amd/compiler/aco_instruction_selection_setup.cpp +++ b/src/amd/compiler/aco_instruction_selection_setup.cpp @@ -328,6 +328,16 @@ setup_tes_variables(isel_context* ctx, nir_shader* nir) } } +void +setup_ms_variables(isel_context* ctx, nir_shader* nir) +{ + setup_vs_output_info(ctx, nir, &ctx->program->info->ms.outinfo); + + ctx->program->config->lds_size = + DIV_ROUND_UP(nir->info.shared_size, ctx->program->dev.lds_encoding_granule); + assert((ctx->program->config->lds_size * ctx->program->dev.lds_encoding_granule) < (32 * 1024)); +} + void setup_variables(isel_context* ctx, nir_shader* nir) { @@ -335,7 +345,8 @@ setup_variables(isel_context* ctx, nir_shader* nir) case MESA_SHADER_FRAGMENT: { break; } - case MESA_SHADER_COMPUTE: { + case MESA_SHADER_COMPUTE: + case MESA_SHADER_TASK: { ctx->program->config->lds_size = DIV_ROUND_UP(nir->info.shared_size, ctx->program->dev.lds_encoding_granule); break; @@ -355,6 +366,10 @@ setup_variables(isel_context* ctx, nir_shader* nir) setup_tes_variables(ctx, nir); break; } + case MESA_SHADER_MESH: { + setup_ms_variables(ctx, nir); + break; + } default: unreachable("Unhandled shader stage."); } @@ -835,6 +850,8 @@ setup_isel_context(Program* program, unsigned shader_count, struct nir_shader* c break; case MESA_SHADER_FRAGMENT: sw_stage = sw_stage | SWStage::FS; break; case MESA_SHADER_COMPUTE: sw_stage = sw_stage | SWStage::CS; break; + case MESA_SHADER_TASK: sw_stage = sw_stage | SWStage::TS; break; + case MESA_SHADER_MESH: sw_stage = sw_stage | SWStage::MS; break; default: unreachable("Shader stage not implemented"); } } @@ -855,6 +872,10 @@ setup_isel_context(Program* program, unsigned shader_count, struct nir_shader* c hw_stage = HWStage::CS; else if (sw_stage == SWStage::GSCopy) hw_stage = HWStage::VS; + else if (sw_stage == SWStage::TS) + hw_stage = HWStage::CS; /* Task shaders are implemented with compute shaders. */ + else if (sw_stage == SWStage::MS) + hw_stage = HWStage::NGG; /* Mesh shaders only work on NGG and on GFX10.3+. */ else if (sw_stage == SWStage::VS_GS && gfx9_plus && !ngg) hw_stage = HWStage::GS; /* GFX6-9: VS+GS merged into a GS (and GFX10/legacy) */ else if (sw_stage == SWStage::VS_GS && ngg) @@ -890,6 +911,10 @@ setup_isel_context(Program* program, unsigned shader_count, struct nir_shader* c program->workgroup_size = program->info->workgroup_size; assert(program->workgroup_size); + /* Mesh shading only works on GFX10.3+. */ + ASSERTED bool mesh_shading = ctx.stage.has(SWStage::TS) || ctx.stage.has(SWStage::MS); + assert(!mesh_shading || ctx.program->chip_class >= GFX10_3); + if (ctx.stage == tess_control_hs) setup_tcs_info(&ctx, shaders[0], NULL); else if (ctx.stage == vertex_tess_control_hs) diff --git a/src/amd/compiler/aco_ir.h b/src/amd/compiler/aco_ir.h index efab59f899d..e8c3e56c2dd 100644 --- a/src/amd/compiler/aco_ir.h +++ b/src/amd/compiler/aco_ir.h @@ -1904,7 +1904,7 @@ struct Block { /* * Shader stages as provided in Vulkan by the application. Contrast this to HWStage. */ -enum class SWStage : uint8_t { +enum class SWStage : uint16_t { None = 0, VS = 1 << 0, /* Vertex Shader */ GS = 1 << 1, /* Geometry Shader */ @@ -1912,7 +1912,9 @@ enum class SWStage : uint8_t { TES = 1 << 3, /* Tessellation Evaluation aka Domain Shader */ FS = 1 << 4, /* Fragment aka Pixel Shader */ CS = 1 << 5, /* Compute Shader */ - GSCopy = 1 << 6, /* GS Copy Shader (internal) */ + TS = 1 << 6, /* Task Shader */ + MS = 1 << 7, /* Mesh Shader */ + GSCopy = 1 << 8, /* GS Copy Shader (internal) */ /* Stage combinations merged to run on a single HWStage */ VS_GS = VS | GS, @@ -1923,7 +1925,7 @@ enum class SWStage : uint8_t { constexpr SWStage operator|(SWStage a, SWStage b) { - return static_cast<SWStage>(static_cast<uint8_t>(a) | static_cast<uint8_t>(b)); + return static_cast<SWStage>(static_cast<uint16_t>(a) | static_cast<uint16_t>(b)); } /* @@ -1956,10 +1958,10 @@ struct Stage { /* Check if the given SWStage is included */ constexpr bool has(SWStage stage) const { - return (static_cast<uint8_t>(sw) & static_cast<uint8_t>(stage)); + return (static_cast<uint16_t>(sw) & static_cast<uint16_t>(stage)); } - unsigned num_sw_stages() const { return util_bitcount(static_cast<uint8_t>(sw)); } + unsigned num_sw_stages() const { return util_bitcount(static_cast<uint16_t>(sw)); } constexpr bool operator==(const Stage& other) const { return sw == other.sw && hw == other.hw; } @@ -1978,6 +1980,9 @@ static constexpr Stage fragment_fs(HWStage::FS, SWStage::FS); static constexpr Stage compute_cs(HWStage::CS, SWStage::CS); static constexpr Stage tess_eval_vs(HWStage::VS, SWStage::TES); static constexpr Stage gs_copy_vs(HWStage::VS, SWStage::GSCopy); +/* Mesh shading pipeline */ +static constexpr Stage task_cs(HWStage::CS, SWStage::TS); +static constexpr Stage mesh_ngg(HWStage::NGG, SWStage::MS); /* GFX10/NGG */ static constexpr Stage vertex_ngg(HWStage::NGG, SWStage::VS); static constexpr Stage vertex_geometry_ngg(HWStage::NGG, SWStage::VS_GS); diff --git a/src/amd/compiler/aco_print_ir.cpp b/src/amd/compiler/aco_print_ir.cpp index 41938fe8625..750f54a4c22 100644 --- a/src/amd/compiler/aco_print_ir.cpp +++ b/src/amd/compiler/aco_print_ir.cpp @@ -795,6 +795,10 @@ print_stage(Stage stage, FILE* output) fprintf(output, "vertex_geometry_ngg"); else if (stage == tess_eval_geometry_ngg) fprintf(output, "tess_eval_geometry_ngg"); + else if (stage == mesh_ngg) + fprintf(output, "mesh_ngg"); + else if (stage == task_cs) + fprintf(output, "task_cs"); else fprintf(output, "unknown");
