From 14409bf1015af47691f09d2be6afb18cfb999aea Mon Sep 17 00:00:00 2001
From: Mukund Keshava <mkeshava@nvidia.com>
Date: Fri, 30 May 2025 16:23:03 +0530
Subject: Enable LSS hit object test (#7273)

* Enable LSS hit object test

Enabled LSS SER tests now that PR #7211, which added SER support to OptiX,
has been merged.

Ran:  ./build/Debug/bin/slangc.exe tests/cuda/lss-test.slang -target ptx
-Xnvrtc -I"C:/ProgramData/NVIDIA Corporation/OptiX SDK 9.0.0/include"

and confirmed that the HitObject intrinsic is called.

eg:
call (%f15, %f16, %f17, %f18,  %f19, %f20, %f21, %f22),
_optix_hitobject_get_linear_curve_vertex_data, ();

* format code

---------

Co-authored-by: slangbot <186143334+slangbot@users.noreply.github.com>
---
 prelude/slang-cuda-prelude.h | 21 +++++++++++----------
 1 file changed, 11 insertions(+), 10 deletions(-)

(limited to 'prelude')

diff --git a/prelude/slang-cuda-prelude.h b/prelude/slang-cuda-prelude.h
index d1160cdd3..a1d3da082 100644
--- a/prelude/slang-cuda-prelude.h
+++ b/prelude/slang-cuda-prelude.h
@@ -31,7 +31,6 @@
 
 #ifdef SLANG_CUDA_ENABLE_OPTIX
 #include <optix.h>
-#include <optix_device.h>
 #endif
 
 // Define slang offsetof implementation
@@ -3226,28 +3225,30 @@ __forceinline__ __device__ float4 optixGetSpherePositionAndRadius()
 {
     float4 data[1];
     optixGetSphereData(data);
-    return data;
+    return data[0];
 }
 
-__forceinline__ __device__ float4 optixHitObjectGetSpherePositionAndRadius()
+__forceinline__ __device__ float4
+optixHitObjectGetSpherePositionAndRadius(OptixTraversableHandle* Obj)
 {
     float4 data[1];
     optixHitObjectGetSphereData(data);
-    return data;
+    return data[0];
 }
 
-__forceinline__ __device__ Matrix<float, 2, 4> optixGetSpherePositionAndRadius()
+__forceinline__ __device__ Matrix<float, 2, 4> optixGetLssPositionsAndRadii()
 {
     float4 data[2];
     optixGetLinearCurveVertexData(data);
-    return Matrix<float, 2, 4>(data[0], data[1]);
+    return makeMatrix<float, 2, 4>(data[0], data[1]);
 }
 
-__forceinline__ __device__ float2x4 optixHitObjectGetSpherePositionAndRadius()
+__forceinline__ __device__ Matrix<float, 2, 4> optixHitObjectGetLssPositionsAndRadii(
+    OptixTraversableHandle* Obj)
 {
     float4 data[2];
     optixHitObjectGetLinearCurveVertexData(data);
-    return Matrix<float, 2, 4>(data[0], data[1]);
+    return makeMatrix<float, 2, 4>(data[0], data[1]);
 }
 
 __forceinline__ __device__ bool optixIsSphereHit()
@@ -3255,7 +3256,7 @@ __forceinline__ __device__ bool optixIsSphereHit()
     return optixGetPrimitiveType() == OPTIX_PRIMITIVE_TYPE_SPHERE;
 }
 
-__forceinline__ __device__ bool optixHitObjectIsSphereHit()
+__forceinline__ __device__ bool optixHitObjectIsSphereHit(OptixTraversableHandle* Obj)
 {
     return optixGetPrimitiveType(optixHitObjectGetHitKind()) == OPTIX_PRIMITIVE_TYPE_SPHERE;
 }
@@ -3265,7 +3266,7 @@ __forceinline__ __device__ bool optixIsLSSHit()
     return optixGetPrimitiveType() == OPTIX_PRIMITIVE_TYPE_ROUND_LINEAR;
 }
 
-__forceinline__ __device__ bool optixHitObjectIsLSSHit()
+__forceinline__ __device__ bool optixHitObjectIsLSSHit(OptixTraversableHandle* Obj)
 {
     return optixGetPrimitiveType(optixHitObjectGetHitKind()) == OPTIX_PRIMITIVE_TYPE_ROUND_LINEAR;
 }
-- 
cgit v1.2.3