Merge pull request #2176 from Try/msl-intersection-params

MSL: ray-query intersection params

6 files changed, 140 insertions, 7 deletions

diff --git a/reference/opt/shaders-msl/comp/ray-query.spv14.vk.ios.msl24..invalid.comp b/reference/opt/shaders-msl/comp/ray-query.spv14.vk.ios.msl24..invalid.comp
index dde7f47b..6e777935 100644
--- a/reference/opt/shaders-msl/comp/ray-query.spv14.vk.ios.msl24..invalid.comp
+++ b/reference/opt/shaders-msl/comp/ray-query.spv14.vk.ios.msl24..invalid.comp
@@ -1,3 +1,5 @@
+#pragma clang diagnostic ignored "-Wmissing-prototypes"
+
 #include <metal_stdlib>
 #include <simd/simd.h>
 #if __METAL_VERSION__ >= 230
@@ -7,6 +9,30 @@ using namespace metal::raytracing;
 
 using namespace metal;
 
+intersection_params spvMakeIntersectionParams(uint flags)
+{
+    intersection_params ip;
+    if ((flags & 1) != 0)
+        ip.force_opacity(forced_opacity::opaque);
+    if ((flags & 2) != 0)
+        ip.force_opacity(forced_opacity::non_opaque);
+    if ((flags & 4) != 0)
+        ip.accept_any_intersection(true);
+    if ((flags & 16) != 0)
+        ip.set_triangle_cull_mode(triangle_cull_mode::back);
+    if ((flags & 32) != 0)
+        ip.set_triangle_cull_mode(triangle_cull_mode::front);
+    if ((flags & 64) != 0)
+        ip.set_opacity_cull_mode(opacity_cull_mode::opaque);
+    if ((flags & 128) != 0)
+        ip.set_opacity_cull_mode(opacity_cull_mode::non_opaque);
+    if ((flags & 256) != 0)
+        ip.set_geometry_cull_mode(geometry_cull_mode::triangle);
+    if ((flags & 512) != 0)
+        ip.set_geometry_cull_mode(geometry_cull_mode::bounding_box);
+    return ip;
+}
+
 struct Params
 {
     uint ray_flags;
@@ -22,9 +48,9 @@ struct Params
 kernel void main0(constant Params& _18 [[buffer(1)]], raytracing::acceleration_structure<raytracing::instancing> AS0 [[buffer(0)]], raytracing::acceleration_structure<raytracing::instancing> AS1 [[buffer(2)]])
 {
     raytracing::intersection_query<raytracing::instancing, raytracing::triangle_data> q;
-    q.reset(ray(_18.origin, _18.dir, _18.tmin, _18.tmax), AS0, intersection_params());
+    q.reset(ray(_18.origin, _18.dir, _18.tmin, _18.tmax), AS0, spvMakeIntersectionParams(_18.ray_flags));
     raytracing::intersection_query<raytracing::instancing, raytracing::triangle_data> q2[2];
-    q2[1].reset(ray(_18.origin, _18.dir, _18.tmin, _18.tmax), AS1, intersection_params());
+    q2[1].reset(ray(_18.origin, _18.dir, _18.tmin, _18.tmax), AS1, spvMakeIntersectionParams(_18.ray_flags));
     bool _63 = q.next();
     bool res = _63;
     q2[0].abort();
diff --git a/reference/opt/shaders-msl/frag/ray-query-object-in-function.spv14.vk.msl24.frag b/reference/opt/shaders-msl/frag/ray-query-object-in-function.spv14.vk.msl24.frag
index 859ace2c..9e23ab04 100644
--- a/reference/opt/shaders-msl/frag/ray-query-object-in-function.spv14.vk.msl24.frag
+++ b/reference/opt/shaders-msl/frag/ray-query-object-in-function.spv14.vk.msl24.frag
@@ -1,3 +1,5 @@
+#pragma clang diagnostic ignored "-Wmissing-prototypes"
+
 #include <metal_stdlib>
 #include <simd/simd.h>
 #if __METAL_VERSION__ >= 230
@@ -7,6 +9,30 @@ using namespace metal::raytracing;
 
 using namespace metal;
 
+intersection_params spvMakeIntersectionParams(uint flags)
+{
+    intersection_params ip;
+    if ((flags & 1) != 0)
+        ip.force_opacity(forced_opacity::opaque);
+    if ((flags & 2) != 0)
+        ip.force_opacity(forced_opacity::non_opaque);
+    if ((flags & 4) != 0)
+        ip.accept_any_intersection(true);
+    if ((flags & 16) != 0)
+        ip.set_triangle_cull_mode(triangle_cull_mode::back);
+    if ((flags & 32) != 0)
+        ip.set_triangle_cull_mode(triangle_cull_mode::front);
+    if ((flags & 64) != 0)
+        ip.set_opacity_cull_mode(opacity_cull_mode::opaque);
+    if ((flags & 128) != 0)
+        ip.set_opacity_cull_mode(opacity_cull_mode::non_opaque);
+    if ((flags & 256) != 0)
+        ip.set_geometry_cull_mode(geometry_cull_mode::triangle);
+    if ((flags & 512) != 0)
+        ip.set_geometry_cull_mode(geometry_cull_mode::bounding_box);
+    return ip;
+}
+
 struct main0_out
 {
     float4 outColor [[color(0)]];
@@ -21,7 +47,7 @@ fragment main0_out main0(main0_in in [[stage_in]], raytracing::acceleration_stru
 {
     main0_out out = {};
     raytracing::intersection_query<raytracing::instancing, raytracing::triangle_data> rayQuery;
-    rayQuery.reset(ray(float3((in.inPos.xy * 4.0) - float2(2.0), 1.0), float3(0.0, 0.0, -1.0), 0.001000000047497451305389404296875, 2.0), topLevelAS, intersection_params());
+    rayQuery.reset(ray(float3((in.inPos.xy * 4.0) - float2(2.0), 1.0), float3(0.0, 0.0, -1.0), 0.001000000047497451305389404296875, 2.0), topLevelAS, spvMakeIntersectionParams(4u));
     for (;;)
     {
         bool _88 = rayQuery.next();
diff --git a/reference/shaders-msl/comp/ray-query.spv14.vk.ios.msl24..invalid.comp b/reference/shaders-msl/comp/ray-query.spv14.vk.ios.msl24..invalid.comp
index dde7f47b..6e777935 100644
--- a/reference/shaders-msl/comp/ray-query.spv14.vk.ios.msl24..invalid.comp
+++ b/reference/shaders-msl/comp/ray-query.spv14.vk.ios.msl24..invalid.comp
@@ -1,3 +1,5 @@
+#pragma clang diagnostic ignored "-Wmissing-prototypes"
+
 #include <metal_stdlib>
 #include <simd/simd.h>
 #if __METAL_VERSION__ >= 230
@@ -7,6 +9,30 @@ using namespace metal::raytracing;
 
 using namespace metal;
 
+intersection_params spvMakeIntersectionParams(uint flags)
+{
+    intersection_params ip;
+    if ((flags & 1) != 0)
+        ip.force_opacity(forced_opacity::opaque);
+    if ((flags & 2) != 0)
+        ip.force_opacity(forced_opacity::non_opaque);
+    if ((flags & 4) != 0)
+        ip.accept_any_intersection(true);
+    if ((flags & 16) != 0)
+        ip.set_triangle_cull_mode(triangle_cull_mode::back);
+    if ((flags & 32) != 0)
+        ip.set_triangle_cull_mode(triangle_cull_mode::front);
+    if ((flags & 64) != 0)
+        ip.set_opacity_cull_mode(opacity_cull_mode::opaque);
+    if ((flags & 128) != 0)
+        ip.set_opacity_cull_mode(opacity_cull_mode::non_opaque);
+    if ((flags & 256) != 0)
+        ip.set_geometry_cull_mode(geometry_cull_mode::triangle);
+    if ((flags & 512) != 0)
+        ip.set_geometry_cull_mode(geometry_cull_mode::bounding_box);
+    return ip;
+}
+
 struct Params
 {
     uint ray_flags;
@@ -22,9 +48,9 @@ struct Params
 kernel void main0(constant Params& _18 [[buffer(1)]], raytracing::acceleration_structure<raytracing::instancing> AS0 [[buffer(0)]], raytracing::acceleration_structure<raytracing::instancing> AS1 [[buffer(2)]])
 {
     raytracing::intersection_query<raytracing::instancing, raytracing::triangle_data> q;
-    q.reset(ray(_18.origin, _18.dir, _18.tmin, _18.tmax), AS0, intersection_params());
+    q.reset(ray(_18.origin, _18.dir, _18.tmin, _18.tmax), AS0, spvMakeIntersectionParams(_18.ray_flags));
     raytracing::intersection_query<raytracing::instancing, raytracing::triangle_data> q2[2];
-    q2[1].reset(ray(_18.origin, _18.dir, _18.tmin, _18.tmax), AS1, intersection_params());
+    q2[1].reset(ray(_18.origin, _18.dir, _18.tmin, _18.tmax), AS1, spvMakeIntersectionParams(_18.ray_flags));
     bool _63 = q.next();
     bool res = _63;
     q2[0].abort();
diff --git a/reference/shaders-msl/frag/ray-query-object-in-function.spv14.vk.msl24.frag b/reference/shaders-msl/frag/ray-query-object-in-function.spv14.vk.msl24.frag
index 3ab6a471..b584f04d 100644
--- a/reference/shaders-msl/frag/ray-query-object-in-function.spv14.vk.msl24.frag
+++ b/reference/shaders-msl/frag/ray-query-object-in-function.spv14.vk.msl24.frag
@@ -9,6 +9,30 @@ using namespace metal::raytracing;
 
 using namespace metal;
 
+intersection_params spvMakeIntersectionParams(uint flags)
+{
+    intersection_params ip;
+    if ((flags & 1) != 0)
+        ip.force_opacity(forced_opacity::opaque);
+    if ((flags & 2) != 0)
+        ip.force_opacity(forced_opacity::non_opaque);
+    if ((flags & 4) != 0)
+        ip.accept_any_intersection(true);
+    if ((flags & 16) != 0)
+        ip.set_triangle_cull_mode(triangle_cull_mode::back);
+    if ((flags & 32) != 0)
+        ip.set_triangle_cull_mode(triangle_cull_mode::front);
+    if ((flags & 64) != 0)
+        ip.set_opacity_cull_mode(opacity_cull_mode::opaque);
+    if ((flags & 128) != 0)
+        ip.set_opacity_cull_mode(opacity_cull_mode::non_opaque);
+    if ((flags & 256) != 0)
+        ip.set_geometry_cull_mode(geometry_cull_mode::triangle);
+    if ((flags & 512) != 0)
+        ip.set_geometry_cull_mode(geometry_cull_mode::bounding_box);
+    return ip;
+}
+
 struct main0_out
 {
     float4 outColor [[color(0)]];
@@ -22,7 +46,7 @@ struct main0_in
 static inline __attribute__((always_inline))
 uint doRay(thread const float3& rayOrigin, thread const float3& rayDirection, thread const float& rayDistance, thread raytracing::intersection_query<raytracing::instancing, raytracing::triangle_data>& rayQuery, thread const raytracing::acceleration_structure<raytracing::instancing>& topLevelAS)
 {
-    rayQuery.reset(ray(rayOrigin, rayDirection, 0.001000000047497451305389404296875, rayDistance), topLevelAS, intersection_params());
+    rayQuery.reset(ray(rayOrigin, rayDirection, 0.001000000047497451305389404296875, rayDistance), topLevelAS, spvMakeIntersectionParams(4u));
     for (;;)
     {
         bool _36 = rayQuery.next();
diff --git a/spirv_msl.cpp b/spirv_msl.cpp
index 791e19cd..a54fb95f 100644
--- a/spirv_msl.cpp
+++ b/spirv_msl.cpp
@@ -7187,6 +7187,35 @@ void CompilerMSL::emit_custom_functions()
 			end_scope();
 			end_scope_decl();
 			statement("");
+			break;
+
+		case SPVFuncImplRayQueryIntersectionParams:
+			statement("intersection_params spvMakeIntersectionParams(uint flags)");
+			begin_scope();
+			statement("intersection_params ip;");
+			statement("if ((flags & ", RayFlagsOpaqueKHRMask, ") != 0)");
+			statement("    ip.force_opacity(forced_opacity::opaque);");
+			statement("if ((flags & ", RayFlagsNoOpaqueKHRMask, ") != 0)");
+			statement("    ip.force_opacity(forced_opacity::non_opaque);");
+			statement("if ((flags & ", RayFlagsTerminateOnFirstHitKHRMask, ") != 0)");
+			statement("    ip.accept_any_intersection(true);");
+			// RayFlagsSkipClosestHitShaderKHRMask is not available in MSL
+			statement("if ((flags & ", RayFlagsCullBackFacingTrianglesKHRMask, ") != 0)");
+			statement("    ip.set_triangle_cull_mode(triangle_cull_mode::back);");
+			statement("if ((flags & ", RayFlagsCullFrontFacingTrianglesKHRMask, ") != 0)");
+			statement("    ip.set_triangle_cull_mode(triangle_cull_mode::front);");
+			statement("if ((flags & ", RayFlagsCullOpaqueKHRMask, ") != 0)");
+			statement("    ip.set_opacity_cull_mode(opacity_cull_mode::opaque);");
+			statement("if ((flags & ", RayFlagsCullNoOpaqueKHRMask, ") != 0)");
+			statement("    ip.set_opacity_cull_mode(opacity_cull_mode::non_opaque);");
+			statement("if ((flags & ", RayFlagsSkipTrianglesKHRMask, ") != 0)");
+			statement("    ip.set_geometry_cull_mode(geometry_cull_mode::triangle);");
+			statement("if ((flags & ", RayFlagsSkipAABBsKHRMask, ") != 0)");
+			statement("    ip.set_geometry_cull_mode(geometry_cull_mode::bounding_box);");
+			statement("return ip;");
+			end_scope();
+			statement("");
+			break;
 
 		default:
 			break;
@@ -9237,10 +9266,11 @@ void CompilerMSL::emit_instruction(const Instruction &instruction)
 	case OpRayQueryInitializeKHR:
 	{
 		flush_variable_declaration(ops[0]);
+		add_spv_func_and_recompile(SPVFuncImplRayQueryIntersectionParams);
 
 		statement(to_expression(ops[0]), ".reset(", "ray(", to_expression(ops[4]), ", ", to_expression(ops[6]), ", ",
 		          to_expression(ops[5]), ", ", to_expression(ops[7]), "), ", to_expression(ops[1]),
-		          ", intersection_params());");
+		          ", spvMakeIntersectionParams(", to_expression(ops[2]), "));");
 		break;
 	}
 	case OpRayQueryProceedKHR:
diff --git a/spirv_msl.hpp b/spirv_msl.hpp
index 4c5c753d..57993978 100644
--- a/spirv_msl.hpp
+++ b/spirv_msl.hpp
@@ -795,6 +795,7 @@ protected:
 		SPVFuncImplConvertYCbCrBT601,
 		SPVFuncImplConvertYCbCrBT2020,
 		SPVFuncImplDynamicImageSampler,
+		SPVFuncImplRayQueryIntersectionParams,
 	};
 
 	// If the underlying resource has been used for comparison then duplicate loads of that resource must be too