From 395302d2404e3429f3cdfa406e89fa76bc0d444b Mon Sep 17 00:00:00 2001 From: Darren Wihandi <65404740+fairywreath@users.noreply.github.com> Date: Thu, 13 Mar 2025 06:28:03 -0400 Subject: Add mesh shader output topology checks (#6592) * initial wip * more wip * add test * add unexpected for invalid target * fixups and improve error message * fixups and improve error message * remove incorrect comment --------- Co-authored-by: Ellie Hermaszewska --- source/slang/slang-diagnostic-defs.h | 6 + source/slang/slang-emit-glsl.cpp | 26 +++- source/slang/slang-emit-hlsl.cpp | 3 - source/slang/slang-emit-metal.cpp | 9 +- source/slang/slang-emit-spirv.cpp | 13 +- source/slang/slang-emit.cpp | 3 + source/slang/slang-ir-entry-point-decorations.cpp | 140 +++++++++++++++++++++ source/slang/slang-ir-entry-point-decorations.h | 36 ++++++ source/slang/slang-ir-inst-defs.h | 2 +- source/slang/slang-ir-insts.h | 2 + source/slang/slang-ir-legalize-varying-params.cpp | 21 +--- source/slang/slang-lower-to-ir.cpp | 14 ++- .../mesh-shader-invalid-output-topology.slang | 48 +++++++ 13 files changed, 283 insertions(+), 40 deletions(-) create mode 100644 source/slang/slang-ir-entry-point-decorations.cpp create mode 100644 source/slang/slang-ir-entry-point-decorations.h create mode 100644 tests/diagnostics/mesh-shader-invalid-output-topology.slang diff --git a/source/slang/slang-diagnostic-defs.h b/source/slang/slang-diagnostic-defs.h index 85ee545a4..4ad942d84 100644 --- a/source/slang/slang-diagnostic-defs.h +++ b/source/slang/slang-diagnostic-defs.h @@ -2437,6 +2437,12 @@ DIAGNOSTIC( invalidTessellationDomain, "'Domain' should be either 'triangles' or 'quads'.") +DIAGNOSTIC( + 50060, + Error, + invalidMeshStageOutputTopology, + "Invalid mesh stage output topology '$0' for target '$1', must be one of: $2") + DIAGNOSTIC( 50082, Error, diff --git a/source/slang/slang-emit-glsl.cpp b/source/slang/slang-emit-glsl.cpp index 86699df48..ac98a0a3c 100644 --- a/source/slang/slang-emit-glsl.cpp +++ b/source/slang/slang-emit-glsl.cpp @@ -4,6 +4,7 @@ #include "../core/slang-writer.h" #include "slang-emit-source-writer.h" #include "slang-ir-call-graph.h" +#include "slang-ir-entry-point-decorations.h" #include "slang-ir-layout.h" #include "slang-ir-util.h" #include "slang-legalize-types.h" @@ -1415,6 +1416,23 @@ void GLSLSourceEmitter::emitParameterGroupImpl( _emitGLSLParameterGroup(varDecl, type); } +static String getOutputTopologyString(OutputTopologyType topology) +{ + SLANG_ASSERT(topology != OutputTopologyType::Unknown); + + switch (topology) + { + case OutputTopologyType::Point: + return "points"; + case OutputTopologyType::Line: + return "lines"; + case OutputTopologyType::Triangle: + return "triangles"; + default: + return ""; + } +} + void GLSLSourceEmitter::emitEntryPointAttributesImpl( IRFunc* irFunc, IREntryPointDecoration* entryPointDecor) @@ -1617,12 +1635,10 @@ void GLSLSourceEmitter::emitEntryPointAttributesImpl( } if (auto decor = as(decoration)) { - // TODO: Ellie validate here/elsewhere, what's allowed here is - // different from the tesselator - // The naming here is plural, so add an 's' m_writer->emit("layout("); - m_writer->emit(decor->getTopology()->getStringSlice()); - m_writer->emit("s) out;\n"); + m_writer->emit( + getOutputTopologyString(OutputTopologyType(decor->getTopologyType()))); + m_writer->emit(") out;\n"); } break; default: diff --git a/source/slang/slang-emit-hlsl.cpp b/source/slang/slang-emit-hlsl.cpp index 89300e13e..0f1ef3ee0 100644 --- a/source/slang/slang-emit-hlsl.cpp +++ b/source/slang/slang-emit-hlsl.cpp @@ -577,9 +577,6 @@ void HLSLSourceEmitter::emitEntryPointAttributesImpl( emitNumThreadsAttribute(); if (auto decor = irFunc->findDecoration()) { - // TODO: Ellie validate here/elsewhere, what's allowed here is - // different from the tesselator - // The naming here is plural, so add an 's' _emitHLSLDecorationSingleString("outputtopology", irFunc, decor->getTopology()); } break; diff --git a/source/slang/slang-emit-metal.cpp b/source/slang/slang-emit-metal.cpp index 1bb738346..37f224083 100644 --- a/source/slang/slang-emit-metal.cpp +++ b/source/slang/slang-emit-metal.cpp @@ -3,6 +3,7 @@ #include "../core/slang-writer.h" #include "slang-emit-source-writer.h" +#include "slang-ir-entry-point-decorations.h" #include "slang-ir-util.h" #include "slang-mangled-lexer.h" @@ -1261,15 +1262,15 @@ void MetalSourceEmitter::emitSimpleTypeImpl(IRType* type) m_writer->emit(", "); emitOperand(meshType->getNumPrimitives(), getInfo(EmitOp::General)); m_writer->emit(", metal::topology::"); - switch (meshType->getTopology()->getValue()) + switch (OutputTopologyType(meshType->getTopology()->getValue())) { - case 1: + case OutputTopologyType::Point: m_writer->emit("point"); break; - case 2: + case OutputTopologyType::Line: m_writer->emit("line"); break; - case 3: + case OutputTopologyType::Triangle: m_writer->emit("triangle"); break; } diff --git a/source/slang/slang-emit-spirv.cpp b/source/slang/slang-emit-spirv.cpp index 3b8e5e1d1..d43af8ac2 100644 --- a/source/slang/slang-emit-spirv.cpp +++ b/source/slang/slang-emit-spirv.cpp @@ -4,6 +4,7 @@ #include "slang-compiler.h" #include "slang-emit-base.h" #include "slang-ir-call-graph.h" +#include "slang-ir-entry-point-decorations.h" #include "slang-ir-insts.h" #include "slang-ir-layout.h" #include "slang-ir-redundancy-removal.h" @@ -4684,7 +4685,7 @@ struct SPIRVEmitContext : public SourceEmitterBase, public SPIRVEmitSharedContex entryPoint ? entryPoint->findDecoration() : nullptr; const auto o = cast(decoration); - const auto t = o->getTopology()->getStringSlice(); + const auto topologyType = OutputTopologyType(o->getTopologyType()); SpvExecutionMode m = SpvExecutionModeMax; if (entryPointDecor) @@ -4693,20 +4694,20 @@ struct SPIRVEmitContext : public SourceEmitterBase, public SPIRVEmitSharedContex { case Stage::Domain: case Stage::Hull: - if (t == "triangle_cw") + if (topologyType == OutputTopologyType::TriangleCW) m = SpvExecutionModeVertexOrderCw; - else if (t == "triangle_ccw") + else if (topologyType == OutputTopologyType::TriangleCCW) m = SpvExecutionModeVertexOrderCcw; break; } } if (m == SpvExecutionModeMax) { - if (t == "triangle") + if (topologyType == OutputTopologyType::Triangle) m = SpvExecutionModeOutputTrianglesEXT; - else if (t == "line") + else if (topologyType == OutputTopologyType::Line) m = SpvExecutionModeOutputLinesEXT; - else if (t == "point") + else if (topologyType == OutputTopologyType::Point) m = SpvExecutionModeOutputPoints; } diff --git a/source/slang/slang-emit.cpp b/source/slang/slang-emit.cpp index 4fb33ccc2..bc113e0c1 100644 --- a/source/slang/slang-emit.cpp +++ b/source/slang/slang-emit.cpp @@ -38,6 +38,7 @@ #include "slang-ir-early-raytracing-intrinsic-simplification.h" #include "slang-ir-eliminate-multilevel-break.h" #include "slang-ir-eliminate-phis.h" +#include "slang-ir-entry-point-decorations.h" #include "slang-ir-entry-point-raw-ptr-params.h" #include "slang-ir-entry-point-uniforms.h" #include "slang-ir-explicit-global-context.h" @@ -722,6 +723,8 @@ Result linkAndOptimizeIR( #endif validateIRModuleIfEnabled(codeGenContext, irModule); + checkEntryPointDecorations(irModule, target, sink); + // Another transformation that needed to wait until we // had layout information on parameters is to take uniform // parameters of a shader entry point and move them into diff --git a/source/slang/slang-ir-entry-point-decorations.cpp b/source/slang/slang-ir-entry-point-decorations.cpp new file mode 100644 index 000000000..eceb94e16 --- /dev/null +++ b/source/slang/slang-ir-entry-point-decorations.cpp @@ -0,0 +1,140 @@ +#include "slang-ir-entry-point-decorations.h" + +#include "compiler-core/slang-diagnostic-sink.h" +#include "core/slang-signal.h" +#include "core/slang-string.h" +#include "core/slang-type-text-util.h" +#include "slang-compiler.h" +#include "slang-ir-insts.h" +#include "slang-ir.h" +#include "slang-options.h" + +namespace Slang +{ + +class CheckEntryPointDecorationsContext +{ +public: + CheckEntryPointDecorationsContext(IRModule* module, CodeGenTarget target, DiagnosticSink* sink) + : m_module(module), m_target(target), m_sink(sink) + { + } + + void check() + { + for (auto inst : m_module->getGlobalInsts()) + { + const auto func = as(inst); + if (!func) + continue; + const auto entryPointDecoration = func->findDecoration(); + if (!entryPointDecoration) + continue; + + checkEntryPoint(func, entryPointDecoration->getProfile().getStage()); + } + } + +private: + void checkEntryPoint(IRFunc* entryPoint, Stage stage) + { + for (auto decoration : entryPoint->getDecorations()) + { + if (auto outputTopologyDecoration = as(decoration)) + { + checkOutputTopologyDecoration(outputTopologyDecoration, stage); + } + } + } + + void checkOutputTopologyDecoration(IROutputTopologyDecoration* decoration, Stage stage) + { + if (stage == Stage::Mesh) + { + const auto outputTopologyType = OutputTopologyType(decoration->getTopologyType()); + if (isTargetGLSL() || isTargetSPIRV() || isTargetMetal()) + { + if (outputTopologyType != OutputTopologyType::Point && + outputTopologyType != OutputTopologyType::Line && + outputTopologyType != OutputTopologyType::Triangle) + { + diagnoseInvalidMeshStageOutputTopology( + decoration, + "'point', 'line', 'triangle'"); + } + } + else if (isTargetHLSL()) + { + if (outputTopologyType != OutputTopologyType::Line && + outputTopologyType != OutputTopologyType::Triangle) + { + diagnoseInvalidMeshStageOutputTopology(decoration, "'line', 'triangle'"); + } + } + else + { + SLANG_UNEXPECTED("Invalid compilation target for mesh stage"); + } + } + } + + void diagnoseInvalidMeshStageOutputTopology( + IROutputTopologyDecoration* decoration, + String validTopologies) + { + m_sink->diagnose( + decoration, + Diagnostics::invalidMeshStageOutputTopology, + decoration->getTopology()->getStringSlice(), + TypeTextUtil::getCompileTargetName(SlangCompileTarget(m_target)), + validTopologies); + } + + bool isTargetHLSL() const { return m_target == CodeGenTarget::HLSL; } + + bool isTargetGLSL() const { return m_target == CodeGenTarget::GLSL; } + + bool isTargetSPIRV() const + { + return m_target == CodeGenTarget::SPIRV || m_target == CodeGenTarget::SPIRVAssembly; + } + + bool isTargetMetal() const + { + return m_target == CodeGenTarget::Metal || m_target == CodeGenTarget::MetalLib || + m_target == CodeGenTarget::MetalLibAssembly; + } + + IRModule* m_module; + const CodeGenTarget m_target; + DiagnosticSink* m_sink; +}; + +void checkEntryPointDecorations(IRModule* module, CodeGenTarget target, DiagnosticSink* sink) +{ + CheckEntryPointDecorationsContext(module, target, sink).check(); +} + +OutputTopologyType convertOutputTopologyStringToEnum(String rawOutputTopology) +{ + auto name = rawOutputTopology.toLower(); + + OutputTopologyType outputTopologyType = OutputTopologyType::Unknown; + +#define CASE(ID, NAME) \ + if (name == String(#NAME).toLower()) \ + { \ + outputTopologyType = OutputTopologyType::ID; \ + } \ + else + + OUTPUT_TOPOLOGY_TYPES(CASE) +#undef CASE + { + outputTopologyType = OutputTopologyType::Unknown; + // no match + } + return outputTopologyType; +} + +} // namespace Slang diff --git a/source/slang/slang-ir-entry-point-decorations.h b/source/slang/slang-ir-entry-point-decorations.h new file mode 100644 index 000000000..87f9ea088 --- /dev/null +++ b/source/slang/slang-ir-entry-point-decorations.h @@ -0,0 +1,36 @@ +// slang-ir-entry-point-decorations.h +#pragma once + +#include "slang-ir.h" + +namespace Slang +{ +enum class CodeGenTarget; +class DiagnosticSink; + +/// Checks entry point decoration values to ensure that they are valid for +/// the shader stage and target. +void checkEntryPointDecorations(IRModule* module, CodeGenTarget target, DiagnosticSink* sink); + + +// OutputTopologyType member definition macro +#define OUTPUT_TOPOLOGY_TYPES(M) \ + M(Point, point) \ + M(Line, line) \ + M(Triangle, triangle) \ + M(TriangleCW, triangle_cw) \ + M(TriangleCCW, triangle_ccw) \ + /* end */ + +enum class OutputTopologyType +{ + Unknown = 0, +#define CASE(ID, NAME) ID, + OUTPUT_TOPOLOGY_TYPES(CASE) +#undef CASE +}; + + +OutputTopologyType convertOutputTopologyStringToEnum(String rawOutputTopology); + +} // namespace Slang diff --git a/source/slang/slang-ir-inst-defs.h b/source/slang/slang-ir-inst-defs.h index d0c0b4b31..574f45243 100644 --- a/source/slang/slang-ir-inst-defs.h +++ b/source/slang/slang-ir-inst-defs.h @@ -860,7 +860,7 @@ INST_RANGE(BindingQuery, GetRegisterIndex, GetRegisterSpace) INST(PatchConstantFuncDecoration, patchConstantFunc, 1, 0) INST(MaxTessFactorDecoration, maxTessFactor, 1, 0) INST(OutputControlPointsDecoration, outputControlPoints, 1, 0) - INST(OutputTopologyDecoration, outputTopology, 1, 0) + INST(OutputTopologyDecoration, outputTopology, 2, 0) INST(PartitioningDecoration, partioning, 1, 0) INST(DomainDecoration, domain, 1, 0) INST(MaxVertexCountDecoration, maxVertexCount, 1, 0) diff --git a/source/slang/slang-ir-insts.h b/source/slang/slang-ir-insts.h index a8a96f230..fc8788b4d 100644 --- a/source/slang/slang-ir-insts.h +++ b/source/slang/slang-ir-insts.h @@ -561,6 +561,8 @@ struct IROutputTopologyDecoration : IRDecoration IR_LEAF_ISA(OutputTopologyDecoration) IRStringLit* getTopology() { return cast(getOperand(0)); } + + IRIntegerValue getTopologyType() { return cast(getOperand(1))->getValue(); } }; struct IRPartitioningDecoration : IRDecoration diff --git a/source/slang/slang-ir-legalize-varying-params.cpp b/source/slang/slang-ir-legalize-varying-params.cpp index e744969db..ad1e89ede 100644 --- a/source/slang/slang-ir-legalize-varying-params.cpp +++ b/source/slang/slang-ir-legalize-varying-params.cpp @@ -3518,27 +3518,8 @@ protected: SLANG_UNEXPECTED("Mesh shader output decoration missing"); return; } - const auto topology = outputDeco->getTopology(); - const auto topStr = topology->getStringSlice(); - UInt topologyEnum = 0; - if (topStr.caseInsensitiveEquals(toSlice("point"))) - { - topologyEnum = 1; - } - else if (topStr.caseInsensitiveEquals(toSlice("line"))) - { - topologyEnum = 2; - } - else if (topStr.caseInsensitiveEquals(toSlice("triangle"))) - { - topologyEnum = 3; - } - else - { - SLANG_UNEXPECTED("unknown topology"); - return; - } + const auto topologyEnum = outputDeco->getTopologyType(); IRInst* topologyConst = builder.getIntValue(builder.getIntType(), topologyEnum); IRType* vertexType = nullptr; diff --git a/source/slang/slang-lower-to-ir.cpp b/source/slang/slang-lower-to-ir.cpp index e43f99b1e..775986a9a 100644 --- a/source/slang/slang-lower-to-ir.cpp +++ b/source/slang/slang-lower-to-ir.cpp @@ -14,6 +14,7 @@ #include "slang-ir-constexpr.h" #include "slang-ir-dce.h" #include "slang-ir-diff-call.h" +#include "slang-ir-entry-point-decorations.h" #include "slang-ir-inline.h" #include "slang-ir-insert-debug-value-store.h" #include "slang-ir-insts.h" @@ -10673,7 +10674,18 @@ struct DeclLoweringVisitor : DeclVisitor else if (auto outputTopAttr = as(modifier)) { IRStringLit* stringLit = _getStringLitFromAttribute(getBuilder(), outputTopAttr); - getBuilder()->addDecoration(irFunc, kIROp_OutputTopologyDecoration, stringLit); + const auto topologyType = + convertOutputTopologyStringToEnum(stringLit->getStringSlice()); + IRInst* topologyTypeInst = getBuilder()->getIntValue( + getBuilder()->getIntType(), + IRIntegerValue(topologyType)); + + auto outputTopologyDecoration = getBuilder()->addDecoration( + irFunc, + kIROp_OutputTopologyDecoration, + stringLit, + topologyTypeInst); + outputTopologyDecoration->sourceLoc = outputTopAttr->loc; } else if (auto maxTessFactortAttr = as(modifier)) { diff --git a/tests/diagnostics/mesh-shader-invalid-output-topology.slang b/tests/diagnostics/mesh-shader-invalid-output-topology.slang new file mode 100644 index 000000000..62f6a8ca7 --- /dev/null +++ b/tests/diagnostics/mesh-shader-invalid-output-topology.slang @@ -0,0 +1,48 @@ +//DIAGNOSTIC_TEST:SIMPLE(filecheck=CHECK_HLSL): -entry main1 -stage mesh -target hlsl +//DIAGNOSTIC_TEST:SIMPLE(filecheck=CHECK): -entry main1 -stage mesh -target spirv +//DIAGNOSTIC_TEST:SIMPLE(filecheck=CHECK): -entry main1 -stage mesh -target glsl +//DIAGNOSTIC_TEST:SIMPLE(filecheck=CHECK): -entry main1 -stage mesh -target metal + +//DIAGNOSTIC_TEST:SIMPLE(filecheck=CHECK_POINT_HLSL): -entry main2 -stage mesh -target hlsl +//TEST:SIMPLE(filecheck=CHECK_POINT): -entry main2 -stage mesh -target spirv +//TEST:SIMPLE(filecheck=CHECK_POINT): -entry main2 -stage mesh -target glsl +//TEST:SIMPLE(filecheck=CHECK_POINT): -entry main2 -stage mesh -target metal + + +struct TaskPayload { + uint offset; +}; + +struct Output { + float4 position : SV_POSITION; +}; + + +// CHECK_HLSL: 50060: Invalid{{.*}}asdqwe{{.*}}of: 'line', 'triangle' +// CHECK: 50060: Invalid{{.*}}asdqwe{{.*}}of: 'point', 'line', 'triangle' +[numthreads(32, 1, 1)] +[outputtopology("asdqwe")] +[shader("mesh")] +void main1( + uint ThreadIndex: SV_GroupIndex, + uint GroupID: SV_GroupID, + out vertices Output Vertices[64], + out indices uint3 Triangles[124], + in payload TaskPayload Payload +) { +} + +// 'point' is not valid for HLSL only, other targets must compile successfully. +// CHECK_POINT_HLSL: 50060: Invalid{{.*}}point{{.*}}of: 'line', 'triangle' +// CHECK_POINT: main +[numthreads(32, 1, 1)] +[outputtopology("point")] +[shader("mesh")] +void main2( + uint ThreadIndex: SV_GroupIndex, + uint GroupID: SV_GroupID, + out vertices Output Vertices[64], + out indices uint Points[124], + in payload TaskPayload Payload +) { +} -- cgit v1.2.3