Add new capdef for lss intrinsics (#7427)

* Add new capdef for lss intrinsics

Fixes #7426

Raygen shaders need to be supported for only hitobject APIs. So we need
a special capability for that, instead of a common one.

* regenerate command line reference

---------

Co-authored-by: slangbot <186143334+slangbot@users.noreply.github.com>

5 files changed, 33 insertions, 4 deletions

diff --git a/docs/command-line-slangc-reference.md b/docs/command-line-slangc-reference.md
index 7d18799ba..12923c24f 100644
--- a/docs/command-line-slangc-reference.md
+++ b/docs/command-line-slangc-reference.md
@@ -1513,6 +1513,7 @@ A capability describes an optional feature that a target may or may not support.
 * `raytracing_intersection` 
 * `raytracing_anyhit_closesthit` 
 * `raytracing_lss` 
+* `raytracing_lss_ho` 
 * `raytracing_anyhit_closesthit_intersection` 
 * `raytracing_raygen_closesthit_miss` 
 * `raytracing_anyhit_closesthit_intersection_miss` 
diff --git a/docs/user-guide/a3-02-reference-capability-atoms.md b/docs/user-guide/a3-02-reference-capability-atoms.md
index 9740806ea..4ef207ea1 100644
--- a/docs/user-guide/a3-02-reference-capability-atoms.md
+++ b/docs/user-guide/a3-02-reference-capability-atoms.md
@@ -1273,6 +1273,11 @@ Compound Capabilities
 `raytracing_lss`
 > Collection of capabilities for linear swept spheres.
 
+`raytracing_lss_ho`
+> hit object APIs allow raygen shaders, but not the non-hit object APIs. So we have this special
+> capdef specifically for the hitobject variant.
+> Collection of capabilities for linear swept spheres.
+
 `raytracing_anyhit_closesthit_intersection`
 > Collection of capabilities for raytracing with the shader stages of anyhit, closesthit, and intersection.
 
diff --git a/source/slang/hlsl.meta.slang b/source/slang/hlsl.meta.slang
index 3cfa9410b..ec7ff0e0d 100644
--- a/source/slang/hlsl.meta.slang
+++ b/source/slang/hlsl.meta.slang
@@ -20662,7 +20662,7 @@ struct HitObject
 
     [__requiresNVAPI]
     [NonUniformReturn]
-    [require(cuda_hlsl_spirv, raytracing_lss)]
+    [require(cuda_hlsl_spirv, raytracing_lss_ho)]
     float4 GetSpherePositionAndRadius()
     {
         __target_switch
@@ -20687,7 +20687,7 @@ struct HitObject
 
     [__requiresNVAPI]
     [NonUniformReturn]
-    [require(cuda_hlsl_spirv, raytracing_lss)]
+    [require(cuda_hlsl_spirv, raytracing_lss_ho)]
     float2x4 GetLssPositionsAndRadii()
     {
         __target_switch
@@ -20718,7 +20718,7 @@ struct HitObject
 
     [__requiresNVAPI]
     [NonUniformReturn]
-    [require(cuda_hlsl_spirv, raytracing_lss)]
+    [require(cuda_hlsl_spirv, raytracing_lss_ho)]
     bool IsSphereHit()
     {
         __target_switch
@@ -20740,7 +20740,7 @@ struct HitObject
 
     [__requiresNVAPI]
     [NonUniformReturn]
-    [require(cuda_hlsl_spirv, raytracing_lss)]
+    [require(cuda_hlsl_spirv, raytracing_lss_ho)]
     bool IsLssHit()
     {
         __target_switch
diff --git a/source/slang/slang-capabilities.capdef b/source/slang/slang-capabilities.capdef
index 343f89687..312777340 100644
--- a/source/slang/slang-capabilities.capdef
+++ b/source/slang/slang-capabilities.capdef
@@ -2260,6 +2260,12 @@ alias raytracing_anyhit_closesthit = anyhit_closesthit + raytracing;
 /// [Compound]
 alias raytracing_lss = raytracing_anyhit_closesthit | spvRayTracingLinearSweptSpheresGeometryNV;
 
+/// hit object APIs allow raygen shaders, but not the non-hit object APIs. So we have this special
+/// capdef specifically for the hitobject variant.
+/// Collection of capabilities for linear swept spheres.
+/// [Compound]
+alias raytracing_lss_ho = raytracing_lss | raygen;
+
 /// Collection of capabilities for raytracing with the shader stages of anyhit, closesthit, and intersection.
 /// [Compound]
 alias raytracing_anyhit_closesthit_intersection = anyhit_closesthit_intersection + raytracing;
diff --git a/tests/cuda/lss-test.slang b/tests/cuda/lss-test.slang
index f3052cacb..d3f4437de 100644
--- a/tests/cuda/lss-test.slang
+++ b/tests/cuda/lss-test.slang
@@ -10,6 +10,12 @@
 //CHECK: optixHitObjectIsSphereHit
 //CHECK: optixHitObjectIsLSSHit
 
+//CHECK_: __global__ void __raygen__raygenShaderLSS()
+//CHECK: optixHitObjectGetSpherePositionAndRadius
+//CHECK: optixHitObjectGetLssPositionsAndRadii
+//CHECK: optixHitObjectIsSphereHit
+//CHECK: optixHitObjectIsLSSHit
+
 struct RayPayload
 {
     float4 color;
@@ -33,4 +39,15 @@ void closestHitShaderLss(inout RayPayload payload, in BuiltInTriangleIntersectio
     float2x4 lssData = hitObj.GetLssPositionsAndRadii();
     bool isSphereHit = hitObj.IsSphereHit();
     bool isLssHit = hitObj.IsLssHit();
+}
+
+[shader("raygen")]
+void raygenShaderLSS(inout RayPayload payload, in BuiltInTriangleIntersectionAttributes attr)
+{
+    // Test HitObject API functions
+    HitObject hitObj;
+    float4 sphereData = hitObj.GetSpherePositionAndRadius();
+    float2x4 lssData = hitObj.GetLssPositionsAndRadii();
+    bool isSphereHit = hitObj.IsSphereHit();
+    bool isLssHit = hitObj.IsLssHit();
 }
\ No newline at end of file