summaryrefslogtreecommitdiffstats
diff options
context:
space:
mode:
Diffstat
-rw-r--r--source/slang/slang-diagnostic-defs.h6
-rw-r--r--source/slang/slang-emit-glsl.cpp26
-rw-r--r--source/slang/slang-emit-hlsl.cpp3
-rw-r--r--source/slang/slang-emit-metal.cpp9
-rw-r--r--source/slang/slang-emit-spirv.cpp13
-rw-r--r--source/slang/slang-emit.cpp3
-rw-r--r--source/slang/slang-ir-entry-point-decorations.cpp140
-rw-r--r--source/slang/slang-ir-entry-point-decorations.h36
-rw-r--r--source/slang/slang-ir-inst-defs.h2
-rw-r--r--source/slang/slang-ir-insts.h2
-rw-r--r--source/slang/slang-ir-legalize-varying-params.cpp21
-rw-r--r--source/slang/slang-lower-to-ir.cpp14
-rw-r--r--tests/diagnostics/mesh-shader-invalid-output-topology.slang48
13 files changed, 283 insertions, 40 deletions
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
@@ -2438,6 +2438,12 @@ DIAGNOSTIC(
"'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,
importingFromPackedBufferUnsupported,
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<IROutputTopologyDecoration>(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<IROutputTopologyDecoration>())
{
- // 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<IREntryPointDecoration>() : nullptr;
const auto o = cast<IROutputTopologyDecoration>(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<IRFunc>(inst);
+ if (!func)
+ continue;
+ const auto entryPointDecoration = func->findDecoration<IREntryPointDecoration>();
+ if (!entryPointDecoration)
+ continue;
+
+ checkEntryPoint(func, entryPointDecoration->getProfile().getStage());
+ }
+ }
+
+private:
+ void checkEntryPoint(IRFunc* entryPoint, Stage stage)
+ {
+ for (auto decoration : entryPoint->getDecorations())
+ {
+ if (auto outputTopologyDecoration = as<IROutputTopologyDecoration>(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<IRStringLit>(getOperand(0)); }
+
+ IRIntegerValue getTopologyType() { return cast<IRIntLit>(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<DeclLoweringVisitor, LoweredValInfo>
else if (auto outputTopAttr = as<OutputTopologyAttribute>(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<MaxTessFactorAttribute>(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
+) {
+}
generated by cgit v1.2.3 (git 2.39.1) at 2026-06-01 14:55:53 +0000