Support Metal 2.4 Intersection Query, Implement GL_EXT_ray_query.

1 files changed, 91 insertions, 0 deletions

diff --git a/reference/opt/shaders-msl/comp/ray-query.nocompat.spv14.vk.comp b/reference/opt/shaders-msl/comp/ray-query.nocompat.spv14.vk.comp
new file mode 100644
index 00000000..b03d524c
--- /dev/null
+++ b/reference/opt/shaders-msl/comp/ray-query.nocompat.spv14.vk.comp
@@ -0,0 +1,91 @@
+#pragma clang diagnostic ignored "-Wmissing-prototypes"
+#pragma clang diagnostic ignored "-Wmissing-braces"
+
+#include <metal_stdlib>
+#include <simd/simd.h>
+using namespace metal::raytracing;
+
+using namespace metal;
+
+template<typename T, size_t Num>
+struct spvUnsafeArray
+{
+    T elements[Num ? Num : 1];
+    
+    thread T& operator [] (size_t pos) thread
+    {
+        return elements[pos];
+    }
+    constexpr const thread T& operator [] (size_t pos) const thread
+    {
+        return elements[pos];
+    }
+    
+    device T& operator [] (size_t pos) device
+    {
+        return elements[pos];
+    }
+    constexpr const device T& operator [] (size_t pos) const device
+    {
+        return elements[pos];
+    }
+    
+    constexpr const constant T& operator [] (size_t pos) const constant
+    {
+        return elements[pos];
+    }
+    
+    threadgroup T& operator [] (size_t pos) threadgroup
+    {
+        return elements[pos];
+    }
+    constexpr const threadgroup T& operator [] (size_t pos) const threadgroup
+    {
+        return elements[pos];
+    }
+};
+
+struct Params
+{
+    uint ray_flags;
+    uint cull_mask;
+    char _m2_pad[8];
+    packed_float3 origin;
+    float tmin;
+    packed_float3 dir;
+    float tmax;
+    float thit;
+};
+
+kernel void main0(constant Params& _18 [[buffer(1)]], acceleration_structure<instancing> AS0 [[buffer(0)]], acceleration_structure<instancing> AS1 [[buffer(2)]])
+{
+    intersection_query<instancing, triangle_data> q;
+    intersection_params _intersection_params_;
+    q.reset(ray(_18.origin, _18.dir, _18.tmin, _18.tmax), AS0, _intersection_params_);
+    spvUnsafeArray<intersection_query<instancing, triangle_data>, 2> q2;
+    intersection_params _intersection_params_;
+    q2[1].reset(ray(_18.origin, _18.dir, _18.tmin, _18.tmax), AS1, _intersection_params_);
+    bool _63 = q.next();
+    q2[0].abort();
+    q.commit_bounding_box_intersection(_18.thit);
+    _14.commit_triangle_intersection();
+    float _71 = q.get_ray_min_distance();
+    float3 _74 = q.get_world_space_ray_origin();
+    float3 _75 = q.get_world_space_ray_direction();
+    uint _80 = (uint)q2[1].get_committed_intersection_type();
+    uint _83 = (uint)q2[0].get_committed_intersection_type();
+    bool _85 = q2[1].is_candidate_non_opaque_bounding_box();
+    float _87 = q2[1].get_committed_distance();
+    float _89 = q2[1].get_committed_distance();
+    int _92 = q.get_committed_user_instance_id();
+    int _94 = q2[0].get_committed_instance_id();
+    int _96 = q2[1].get_committed_geometry_id();
+    int _97 = q.get_committed_primitive_id();
+    float2 _100 = q2[0].get_committed_triangle_barycentric_coord();
+    bool _103 = q.is_committed_triangle_front_facing();
+    float3 _104 = q.get_committed_ray_direction();
+    float3 _106 = q2[0].get_committed_ray_origin();
+    float4x3 _110 = q.get_committed_object_to_world_transform();
+    float4x3 _112 = q2[1].get_committed_world_to_object_transform();
+}
+