Merge pull request #1783 from billhollings/more-unpacked-vectors
MSL: Support more usecases for unpacked vectors.
diff --git a/reference/opt/shaders-msl/comp/ray-query.spv14.vk.ios.msl24.comp b/reference/opt/shaders-msl/comp/ray-query.spv14.vk.ios.msl24.comp
new file mode 100644
index 0000000..3e86c56
--- /dev/null
+++ b/reference/opt/shaders-msl/comp/ray-query.spv14.vk.ios.msl24.comp
@@ -0,0 +1,51 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+#if __METAL_VERSION__ >= 230
+#include <metal_raytracing>
+using namespace metal::raytracing;
+#endif
+
+using namespace metal;
+
+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;
+ q.reset(ray(_18.origin, _18.dir, _18.tmin, _18.tmax), AS0, intersection_params());
+ intersection_query<instancing, triangle_data> q2[2];
+ 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);
+ q2[1].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_candidate_intersection_type()) - 1;
+ bool _85 = q2[1].is_candidate_non_opaque_bounding_box();
+ float _87 = q2[1].get_committed_distance();
+ float _89 = q2[1].get_candidate_triangle_distance();
+ int _92 = q.get_committed_user_instance_id();
+ int _94 = q2[0].get_candidate_instance_id();
+ int _96 = q2[1].get_candidate_geometry_id();
+ int _97 = q.get_committed_primitive_id();
+ float2 _100 = q2[0].get_candidate_triangle_barycentric_coord();
+ bool _103 = q.is_committed_triangle_front_facing();
+ float3 _104 = q.get_candidate_ray_direction();
+ float3 _106 = q2[0].get_committed_ray_origin();
+ float4x3 _110 = q.get_candidate_object_to_world_transform();
+ float4x3 _112 = q2[1].get_committed_world_to_object_transform();
+}
+
diff --git a/reference/shaders-msl/comp/ray-query.spv14.vk.ios.msl24.comp b/reference/shaders-msl/comp/ray-query.spv14.vk.ios.msl24.comp
index f73d491..fdee9d6 100644
--- a/reference/shaders-msl/comp/ray-query.spv14.vk.ios.msl24.comp
+++ b/reference/shaders-msl/comp/ray-query.spv14.vk.ios.msl24.comp
@@ -1,6 +1,3 @@
-#pragma clang diagnostic ignored "-Wmissing-prototypes"
-#pragma clang diagnostic ignored "-Wmissing-braces"
-
#include <metal_stdlib>
#include <simd/simd.h>
#if __METAL_VERSION__ >= 230
@@ -10,44 +7,6 @@
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;
@@ -64,13 +23,13 @@
{
intersection_query<instancing, triangle_data> q;
q.reset(ray(_18.origin, _18.dir, _18.tmin, _18.tmax), AS0, intersection_params());
- spvUnsafeArray<intersection_query<instancing, triangle_data>, 2> q2;
+ intersection_query<instancing, triangle_data> q2[2];
q2[1].reset(ray(_18.origin, _18.dir, _18.tmin, _18.tmax), AS1, intersection_params());
bool _63 = q.next();
bool res = _63;
q2[0].abort();
q.commit_bounding_box_intersection(_18.thit);
- _14.commit_triangle_intersection();
+ q2[1].commit_triangle_intersection();
float _71 = q.get_ray_min_distance();
float fval = _71;
float3 _74 = q.get_world_space_ray_origin();
diff --git a/spirv_msl.cpp b/spirv_msl.cpp
index c38a76c..43cb6cd 100644
--- a/spirv_msl.cpp
+++ b/spirv_msl.cpp
@@ -8381,10 +8381,8 @@
case OpConvertUToAccelerationStructureKHR:
SPIRV_CROSS_THROW("ConvertUToAccelerationStructure is not supported in MSL.");
- break; // Nothing to do in the body
case OpRayQueryGetIntersectionInstanceShaderBindingTableRecordOffsetKHR:
SPIRV_CROSS_THROW("BindingTableRecordOffset is not supported in MSL.");
- break; // Nothing to do in the body
case OpRayQueryInitializeKHR:
{
@@ -8458,7 +8456,7 @@
}
case OpRayQueryConfirmIntersectionKHR:
flush_variable_declaration(ops[0]);
- statement(to_expression(ops[2]), ".commit_triangle_intersection();");
+ statement(to_expression(ops[0]), ".commit_triangle_intersection();");
break;
case OpRayQueryGenerateIntersectionKHR:
flush_variable_declaration(ops[0]);
@@ -13402,8 +13400,7 @@
SPIRV_CROSS_THROW("Acceleration Structure Type is supported in MSL 2.3 and above.");
break;
case SPIRType::RayQuery:
- type_name = "intersection_query<instancing, triangle_data>";
- break;
+ return "intersection_query<instancing, triangle_data>";
default:
return "unknown_type";
@@ -13448,6 +13445,7 @@
{
case SPIRType::AtomicCounter:
case SPIRType::ControlPointArray:
+ case SPIRType::RayQuery:
{
return CompilerGLSL::type_to_array_glsl(type);
}
