Add LSS intrinsics (#7200)

* WiP: LSS intrinsics: initial commit

* format code

* Fix CI failures

* Address review comment

---------

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

1 files changed, 49 insertions, 0 deletions

diff --git a/prelude/slang-cuda-prelude.h b/prelude/slang-cuda-prelude.h
index d2c9fce9d..d1160cdd3 100644
--- a/prelude/slang-cuda-prelude.h
+++ b/prelude/slang-cuda-prelude.h
@@ -31,6 +31,7 @@
 
 #ifdef SLANG_CUDA_ENABLE_OPTIX
 #include <optix.h>
+#include <optix_device.h>
 #endif
 
 // Define slang offsetof implementation
@@ -3221,6 +3222,54 @@ __forceinline__ __device__ void* optixTrace(
         r1);
 }
 
+__forceinline__ __device__ float4 optixGetSpherePositionAndRadius()
+{
+    float4 data[1];
+    optixGetSphereData(data);
+    return data;
+}
+
+__forceinline__ __device__ float4 optixHitObjectGetSpherePositionAndRadius()
+{
+    float4 data[1];
+    optixHitObjectGetSphereData(data);
+    return data;
+}
+
+__forceinline__ __device__ Matrix<float, 2, 4> optixGetSpherePositionAndRadius()
+{
+    float4 data[2];
+    optixGetLinearCurveVertexData(data);
+    return Matrix<float, 2, 4>(data[0], data[1]);
+}
+
+__forceinline__ __device__ float2x4 optixHitObjectGetSpherePositionAndRadius()
+{
+    float4 data[2];
+    optixHitObjectGetLinearCurveVertexData(data);
+    return Matrix<float, 2, 4>(data[0], data[1]);
+}
+
+__forceinline__ __device__ bool optixIsSphereHit()
+{
+    return optixGetPrimitiveType() == OPTIX_PRIMITIVE_TYPE_SPHERE;
+}
+
+__forceinline__ __device__ bool optixHitObjectIsSphereHit()
+{
+    return optixGetPrimitiveType(optixHitObjectGetHitKind()) == OPTIX_PRIMITIVE_TYPE_SPHERE;
+}
+
+__forceinline__ __device__ bool optixIsLSSHit()
+{
+    return optixGetPrimitiveType() == OPTIX_PRIMITIVE_TYPE_ROUND_LINEAR;
+}
+
+__forceinline__ __device__ bool optixHitObjectIsLSSHit()
+{
+    return optixGetPrimitiveType(optixHitObjectGetHitKind()) == OPTIX_PRIMITIVE_TYPE_ROUND_LINEAR;
+}
+
 template<typename T>
 __forceinline__ __device__ void* optixTraverse(
     OptixTraversableHandle AccelerationStructure,