diff --git a/source/compiler-core/slang-downstream-compiler.h b/source/compiler-core/slang-downstream-compiler.h index 5365b983960..c23a6eff032 100644 --- a/source/compiler-core/slang-downstream-compiler.h +++ b/source/compiler-core/slang-downstream-compiler.h @@ -260,6 +260,9 @@ struct DownstreamCompileOptions /// Profile name to use, only required for compiles that need to compile against a a specific /// profiles. Profile names are tied to compilers and targets. TerminatedCharSlice profileName; + // According to DirectX Raytracing Specification, PAQs are supported in Shader Model 6.7 and + // above + bool enablePAQ = false; /// The stage being compiled for SlangStage stage = SLANG_STAGE_NONE; diff --git a/source/compiler-core/slang-dxc-compiler.cpp b/source/compiler-core/slang-dxc-compiler.cpp index 065ee41451d..0d4bc0a5970 100644 --- a/source/compiler-core/slang-dxc-compiler.cpp +++ b/source/compiler-core/slang-dxc-compiler.cpp @@ -479,8 +479,7 @@ SlangResult DXCDownstreamCompiler::compile(const CompileOptions& inOptions, IArt args.add(compilerSpecific[i]); } - // This can be re-enabled when we add PAQs: https://github.com/shader-slang/slang/issues/3448 - const bool enablePAQs = false; + bool enablePAQs = options.enablePAQ; if (!enablePAQs) args.add(L"-disable-payload-qualifiers"); else diff --git a/source/slang/core.meta.slang b/source/slang/core.meta.slang index e2fb8bbf27f..481aba1918b 100644 --- a/source/slang/core.meta.slang +++ b/source/slang/core.meta.slang @@ -4214,3 +4214,6 @@ attribute_syntax [RequireFullQuads] : RequireFullQuadsAttribute; __generic typealias NodePayloadPtr = Ptr; +__attributeTarget(StructDecl) +attribute_syntax [raypayload] : RayPayloadAttribute; + diff --git a/source/slang/slang-ast-modifier.h b/source/slang/slang-ast-modifier.h index 5f9ccb5bbad..86c1b556c86 100644 --- a/source/slang/slang-ast-modifier.h +++ b/source/slang/slang-ast-modifier.h @@ -1698,6 +1698,16 @@ class PayloadAttribute : public Attribute SLANG_AST_CLASS(PayloadAttribute) }; +/// A `[raypayload]` attribute indicates that a `struct` type will be used as +/// a ray payload for `TraceRay()` calls, and thus also as input/output +/// for shaders in the ray tracing pipeline that might be invoked for +/// such a ray. +/// +class RayPayloadAttribute : public Attribute +{ + SLANG_AST_CLASS(RayPayloadAttribute) +}; + /// A `[deprecated("message")]` attribute indicates the target is /// deprecated. /// A compiler warning including the message will be raised if the diff --git a/source/slang/slang-check-decl.cpp b/source/slang/slang-check-decl.cpp index 4ab90911887..21a16cae535 100644 --- a/source/slang/slang-check-decl.cpp +++ b/source/slang/slang-check-decl.cpp @@ -12496,6 +12496,12 @@ void SemanticsDeclAttributesVisitor::visitStructDecl(StructDecl* structDecl) } } + // Check if this is a ray payload struct and validate field access qualifiers + if (structDecl->findModifier()) + { + checkRayPayloadStructFields(structDecl); + } + int backingWidth = 0; [[maybe_unused]] int totalWidth = 0; struct BitFieldInfo diff --git a/source/slang/slang-check-impl.h b/source/slang/slang-check-impl.h index f7681ba4571..8a1e79ce86a 100644 --- a/source/slang/slang-check-impl.h +++ b/source/slang/slang-check-impl.h @@ -2835,6 +2835,8 @@ struct SemanticsVisitor : public SemanticsContext bool isCStyleType(Type* type, HashSet& isVisit); void addVisibilityModifier(Decl* decl, DeclVisibility vis); + + void checkRayPayloadStructFields(StructDecl* structDecl); }; diff --git a/source/slang/slang-check-modifier.cpp b/source/slang/slang-check-modifier.cpp index 741823a65f1..d94c77d6a71 100644 --- a/source/slang/slang-check-modifier.cpp +++ b/source/slang/slang-check-modifier.cpp @@ -1413,9 +1413,20 @@ bool isModifierAllowedOnDecl(bool isGLSLInput, ASTNodeType modifierType, Decl* d case ASTNodeType::ConstRefModifier: case ASTNodeType::GLSLBufferModifier: case ASTNodeType::GLSLPatchModifier: + return (as(decl) && isGlobalDecl(decl)) || as(decl) || + as(decl); case ASTNodeType::RayPayloadAccessSemantic: case ASTNodeType::RayPayloadReadSemantic: case ASTNodeType::RayPayloadWriteSemantic: + // Allow on struct fields if the parent struct has the [raypayload] attribute + if (auto varDecl = as(decl)) + { + if (auto structDecl = as(varDecl->parentDecl)) + { + if (structDecl->findModifier()) + return true; + } + } return (as(decl) && isGlobalDecl(decl)) || as(decl) || as(decl); @@ -2179,5 +2190,36 @@ void SemanticsVisitor::checkModifiers(ModifiableSyntaxNode* syntaxNode) postProcessingOnModifiers(syntaxNode->modifiers); } +void SemanticsVisitor::checkRayPayloadStructFields(StructDecl* structDecl) +{ + // Only check structs with the [raypayload] attribute + if (!structDecl->findModifier()) + { + return; + } + + // Check each field in the struct + for (auto member : structDecl->members) + { + auto fieldVarDecl = as(member); + if (!fieldVarDecl) + { + continue; + } + + bool hasReadModifier = fieldVarDecl->findModifier() != nullptr; + bool hasWriteModifier = fieldVarDecl->findModifier() != nullptr; + + if (!hasReadModifier && !hasWriteModifier) + { + // Emit the diagnostic error + getSink()->diagnose( + fieldVarDecl, + Diagnostics::rayPayloadFieldMissingAccessQualifiers, + fieldVarDecl->getName()); + } + } +} + } // namespace Slang diff --git a/source/slang/slang-compiler.cpp b/source/slang/slang-compiler.cpp index 3839e072259..55f3846af97 100644 --- a/source/slang/slang-compiler.cpp +++ b/source/slang/slang-compiler.cpp @@ -1732,6 +1732,11 @@ SlangResult CodeGenContext::emitWithDownstreamForEntryPoints(ComPtr& options.libraries = SliceUtil::asSlice(libraries); options.libraryPaths = allocator.allocate(libraryPaths); + if (m_targetProfile.getFamily() == ProfileFamily::DX) + { + options.enablePAQ = m_targetProfile.getVersion() >= ProfileVersion::DX_6_7; + } + // Compile ComPtr artifact; auto downstreamStartTime = std::chrono::high_resolution_clock::now(); diff --git a/source/slang/slang-compiler.h b/source/slang/slang-compiler.h index 18192678a15..8a9b8985a6d 100644 --- a/source/slang/slang-compiler.h +++ b/source/slang/slang-compiler.h @@ -2812,7 +2812,9 @@ struct CodeGenContext }; CodeGenContext(Shared* shared) - : m_shared(shared), m_targetFormat(shared->targetProgram->getTargetReq()->getTarget()) + : m_shared(shared) + , m_targetFormat(shared->targetProgram->getTargetReq()->getTarget()) + , m_targetProfile(shared->targetProgram->getOptionSet().getProfile()) { } @@ -2909,6 +2911,7 @@ struct CodeGenContext protected: CodeGenTarget m_targetFormat = CodeGenTarget::Unknown; + Profile m_targetProfile; ExtensionTracker* m_extensionTracker = nullptr; /// Will output assembly as well as the artifact if appropriate for the artifact type for diff --git a/source/slang/slang-diagnostic-defs.h b/source/slang/slang-diagnostic-defs.h index f2c7fecc1ee..21bf73d6ede 100644 --- a/source/slang/slang-diagnostic-defs.h +++ b/source/slang/slang-diagnostic-defs.h @@ -2702,4 +2702,14 @@ DIAGNOSTIC( noBlocksOrIntrinsic, "no blocks found for function definition, is there a '$0' intrinsic missing?") +// +// Ray tracing +// + +DIAGNOSTIC( + 40000, + Error, + rayPayloadFieldMissingAccessQualifiers, + "field '$0' in ray payload struct must have either 'read' OR 'write' access qualifiers") + #undef DIAGNOSTIC diff --git a/source/slang/slang-emit-hlsl.cpp b/source/slang/slang-emit-hlsl.cpp index 0f1ef3ee099..2d963866d9b 100644 --- a/source/slang/slang-emit-hlsl.cpp +++ b/source/slang/slang-emit-hlsl.cpp @@ -1667,8 +1667,18 @@ void HLSLSourceEmitter::emitPostKeywordTypeAttributesImpl(IRInst* inst) { m_writer->emit("[payload] "); } - // This can be re-enabled when we add PAQs: https://github.com/shader-slang/slang/issues/3448 - const bool enablePAQs = false; + + // Get the target profile to determine if PAQs are supported + bool enablePAQs = false; + auto profile = getTargetProgram()->getOptionSet().getProfile(); + if (profile.getFamily() == ProfileFamily::DX) + { + // PAQs are default in Shader Model 6.7 and above when called with `--profile lib_6_7` + + auto version = profile.getVersion(); + enablePAQs = version >= ProfileVersion::DX_6_7; + } + if (enablePAQs) { if (const auto payloadDecoration = inst->findDecoration()) diff --git a/source/slang/slang-lower-to-ir.cpp b/source/slang/slang-lower-to-ir.cpp index e6ec68660bb..260596dc3bd 100644 --- a/source/slang/slang-lower-to-ir.cpp +++ b/source/slang/slang-lower-to-ir.cpp @@ -9318,6 +9318,11 @@ struct DeclLoweringVisitor : DeclVisitor subBuilder->addDecoration(irAggType, kIROp_PayloadDecoration); } + if (const auto rayPayloadAttribute = decl->findModifier()) + { + subBuilder->addDecoration(irAggType, kIROp_RayPayloadDecoration); + } + subBuilder->setInsertInto(irAggType); // A `struct` that inherits from another `struct` must start diff --git a/tests/diagnostics/raypayload-missing-access-qualifiers.slang b/tests/diagnostics/raypayload-missing-access-qualifiers.slang new file mode 100644 index 00000000000..d22a6300b49 --- /dev/null +++ b/tests/diagnostics/raypayload-missing-access-qualifiers.slang @@ -0,0 +1,37 @@ +// raypayload-missing-access-qualifiers.slang + +//DIAGNOSTIC_TEST:SIMPLE: + +// Test error for field in ray payload struct missing read/write access qualifiers + +struct [raypayload] RayPayload +{ + float4 color : read(caller, anyhit) : write(caller); + float4 colorMissingQualifiers; // Error expected here + +}; + +uniform RWTexture2D resultTexture; +uniform RaytracingAccelerationStructure sceneBVH; + +[shader("raygeneration")] +void rayGenShaderA() +{ + int2 threadIdx = DispatchRaysIndex().xy; + + float3 rayDir = float3(0, 0, 1); + float3 rayOrigin = 0; + rayOrigin.x = (threadIdx.x * 2) - 1; + rayOrigin.y = (threadIdx.y * 2) - 1; + + // Trace the ray. + RayDesc ray; + ray.Origin = rayOrigin; + ray.Direction = rayDir; + ray.TMin = 0.001; + ray.TMax = 10000.0; + RayPayload payload = { float4(0, 0, 0, 0) , float4(0, 0, 0, 0)}; + TraceRay(sceneBVH, RAY_FLAG_NONE, ~0, 0, 0, 0, ray, payload); + + resultTexture[threadIdx.xy] = payload.color; +} diff --git a/tests/diagnostics/raypayload-missing-access-qualifiers.slang.expected b/tests/diagnostics/raypayload-missing-access-qualifiers.slang.expected new file mode 100644 index 00000000000..525e8529cd4 --- /dev/null +++ b/tests/diagnostics/raypayload-missing-access-qualifiers.slang.expected @@ -0,0 +1,8 @@ +result code = -1 +standard error = { +tests/diagnostics/raypayload-missing-access-qualifiers.slang(10): error 40000: field 'colorMissingQualifiers' in ray payload struct must have either 'read' OR 'write' access qualifiers + float4 colorMissingQualifiers; // Error expected here + ^~~~~~~~~~~~~~~~~~~~~~ +} +standard output = { +} diff --git a/tests/hlsl/raypayload-attribute-no-struct.slang b/tests/hlsl/raypayload-attribute-no-struct.slang index c7ad9459354..4e4921e14bb 100644 --- a/tests/hlsl/raypayload-attribute-no-struct.slang +++ b/tests/hlsl/raypayload-attribute-no-struct.slang @@ -1,7 +1,8 @@ -//enable when https://github.com/shader-slang/slang/issues/3448 is implemented -//DISABLE_TEST:SIMPLE(filecheck=CHECK): -target hlsl -stage raygeneration -entry rayGenShaderA +//TEST:SIMPLE(filecheck=CHECK): -target hlsl -profile lib_6_6 -stage raygeneration -entry rayGenShaderA +//TEST:SIMPLE(filecheck=DXIL): -target dxil -profile lib_6_6 -stage raygeneration -entry rayGenShaderA -// CHECK: struct [raypayload] +// CHECK: struct RayPayload +// DXIL: define void @ uniform RWTexture2D resultTexture; uniform RaytracingAccelerationStructure sceneBVH; diff --git a/tests/hlsl/raypayload-attribute-paq.slang b/tests/hlsl/raypayload-attribute-paq.slang new file mode 100644 index 00000000000..3af0556bc0b --- /dev/null +++ b/tests/hlsl/raypayload-attribute-paq.slang @@ -0,0 +1,37 @@ +//TEST:SIMPLE(filecheck=CHECK): -target hlsl -profile lib_6_7 -stage raygeneration -entry rayGenShaderA +//TEST:SIMPLE(filecheck=DXIL): -target dxil -profile lib_6_7 -stage raygeneration -entry rayGenShaderA + +// CHECK: struct [raypayload] +// CHECK: float4 color_0 : read(caller, anyhit) : write(caller); +// DXIL: define void @ +// DXIL: !dx.dxrPayloadAnnotations + +struct [raypayload] RayPayload +{ + float4 color : read(caller, anyhit) : write(caller); +}; + +uniform RWTexture2D resultTexture; +uniform RaytracingAccelerationStructure sceneBVH; + +[shader("raygeneration")] +void rayGenShaderA() +{ + int2 threadIdx = DispatchRaysIndex().xy; + + float3 rayDir = float3(0, 0, 1); + float3 rayOrigin = 0; + rayOrigin.x = (threadIdx.x * 2) - 1; + rayOrigin.y = (threadIdx.y * 2) - 1; + + // Trace the ray. + RayDesc ray; + ray.Origin = rayOrigin; + ray.Direction = rayDir; + ray.TMin = 0.001; + ray.TMax = 10000.0; + RayPayload payload = { float4(0, 0, 0, 0) }; + TraceRay(sceneBVH, RAY_FLAG_NONE, ~0, 0, 0, 0, ray, payload); + + resultTexture[threadIdx.xy] = payload.color; +} diff --git a/tests/hlsl/raypayload-attribute.slang b/tests/hlsl/raypayload-attribute.slang index b981589ac81..1a9e9a7f53b 100644 --- a/tests/hlsl/raypayload-attribute.slang +++ b/tests/hlsl/raypayload-attribute.slang @@ -1,8 +1,9 @@ -//enable when https://github.com/shader-slang/slang/issues/3448 is implemented -//DISABLE_TEST:SIMPLE(filecheck=CHECK): -target hlsl -stage raygeneration -entry rayGenShaderA - -// CHECK: struct [raypayload] +//TEST:SIMPLE(filecheck=CHECK): -target hlsl -profile lib_6_6 -stage raygeneration -entry rayGenShaderA +//TEST:SIMPLE(filecheck=DXIL): -target dxil -profile lib_6_6 -stage raygeneration -entry rayGenShaderA +// CHECK: struct RayPayload +// CHECK: float4 color +// DXIL: define void @ struct RayPayload { float4 color; diff --git a/tools/gfx-unit-test/ray-tracing-test-shaders.slang b/tools/gfx-unit-test/ray-tracing-test-shaders.slang index aa2e5055fc2..c1273a717e4 100644 --- a/tools/gfx-unit-test/ray-tracing-test-shaders.slang +++ b/tools/gfx-unit-test/ray-tracing-test-shaders.slang @@ -1,8 +1,8 @@ // ray-tracing-test-shaders.slang -struct RayPayload +struct [raypayload] RayPayload { - float4 color; + float4 color : read(caller) : write(caller, closesthit, miss); }; uniform RWTexture2D resultTexture;