Skip to content

Commit

Permalink
Merge pull request KhronosGroup#2176 from Try/msl-intersection-params
Browse files Browse the repository at this point in the history 
MSL: ray-query intersection params
  • Loading branch information
@HansKristian-Work
HansKristian-Work authored Jul 14, 2023
2 parents b8e742c + 6b2ae11 commit b43c1a1
Show file tree
Hide file tree
Showing 6 changed files with 140 additions and 7 deletions.
30 changes: 28 additions & 2 deletions reference/opt/shaders-msl/comp/ray-query.spv14.vk.ios.msl24..invalid.comp
View file
Original file line number Diff line number Diff line change
@@ -1,3 +1,5 @@
#pragma clang diagnostic ignored "-Wmissing-prototypes"

#include <metal_stdlib>
#include <simd/simd.h>
#if __METAL_VERSION__ >= 230
Expand All @@ -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;
Expand All @@ -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();
Expand Down
28 changes: 27 additions & 1 deletion reference/opt/shaders-msl/frag/ray-query-object-in-function.spv14.vk.msl24.frag
View file
Original file line number Diff line number Diff line change
@@ -1,3 +1,5 @@
#pragma clang diagnostic ignored "-Wmissing-prototypes"

#include <metal_stdlib>
#include <simd/simd.h>
#if __METAL_VERSION__ >= 230
Expand All @@ -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)]];
Expand All @@ -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();
Expand Down
30 changes: 28 additions & 2 deletions reference/shaders-msl/comp/ray-query.spv14.vk.ios.msl24..invalid.comp
View file
Original file line number Diff line number Diff line change
@@ -1,3 +1,5 @@
#pragma clang diagnostic ignored "-Wmissing-prototypes"

#include <metal_stdlib>
#include <simd/simd.h>
#if __METAL_VERSION__ >= 230
Expand All @@ -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;
Expand All @@ -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();
Expand Down
26 changes: 25 additions & 1 deletion reference/shaders-msl/frag/ray-query-object-in-function.spv14.vk.msl24.frag
View file
Original file line number Diff line number Diff line change
Expand Up @@ -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)]];
Expand All @@ -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();
Expand Down
32 changes: 31 additions & 1 deletion spirv_msl.cpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -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;
Expand Down Expand Up @@ -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:
Expand Down
1 change: 1 addition & 0 deletions spirv_msl.hpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -795,6 +795,7 @@ class CompilerMSL : public CompilerGLSL
SPVFuncImplConvertYCbCrBT601,
SPVFuncImplConvertYCbCrBT2020,
SPVFuncImplDynamicImageSampler,
SPVFuncImplRayQueryIntersectionParams,
};

// If the underlying resource has been used for comparison then duplicate loads of that resource must be too
Expand Down

0 comments on commit b43c1a1

Please sign in to comment.