Mesa (main): aco: Add Mesh and Task shader stages.
GitLab Mirror
gitlab-mirror at kemper.freedesktop.org
Fri Dec 31 13:38:48 UTC 2021
Module: Mesa
Branch: main
Commit: 6766e6a985d7a1ea3dc0eb3896c9fce745b9717b
URL: http://cgit.freedesktop.org/mesa/mesa/commit/?id=6766e6a985d7a1ea3dc0eb3896c9fce745b9717b
Author: Timur Kristóf <timur.kristof at gmail.com>
Date: Thu Oct 21 11:33:10 2021 +0200
aco: Add Mesh and Task shader stages.
Signed-off-by: Timur Kristóf <timur.kristof at gmail.com>
Reviewed-by: Rhys Perry <pendingchaos02 at gmail.com>
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");
More information about the mesa-commit
mailing list