diff --git a/reference/shaders-msl-no-opt/frag/ray-query-mutability.spv14.vk.msl24.frag b/reference/shaders-msl-no-opt/frag/ray-query-mutability.spv14.vk.msl24.frag
new file mode 100644
index 000000000..720fba1e5
--- /dev/null
+++ b/reference/shaders-msl-no-opt/frag/ray-query-mutability.spv14.vk.msl24.frag
@@ -0,0 +1,70 @@
+#pragma clang diagnostic ignored "-Wmissing-prototypes"
+
+#include <metal_stdlib>
+#include <simd/simd.h>
+#if __METAL_VERSION__ >= 230
+#include <metal_raytracing>
+using namespace metal::raytracing;
+#endif
+
+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;
+}
+
+static inline __attribute__((always_inline))
+void initFn(thread raytracing::intersection_query<raytracing::instancing, raytracing::triangle_data>& rayQuery, const raytracing::acceleration_structure<raytracing::instancing> topLevelAS)
+{
+    float3 rayOrigin = float3(0.0, 0.0, 1.0);
+    float3 rayDirection = float3(0.0, 0.0, -1.0);
+    float rayDistance = 2.0;
+    rayQuery.reset(ray(rayOrigin, rayDirection, 0.001000000047497451305389404296875, rayDistance), topLevelAS, spvMakeIntersectionParams(4u));
+}
+
+static inline __attribute__((always_inline))
+uint proceeFn(thread raytracing::intersection_query<raytracing::instancing, raytracing::triangle_data>& rayQuery)
+{
+    for (;;)
+    {
+        bool _46 = rayQuery.next();
+        if (_46)
+        {
+            continue;
+        }
+        else
+        {
+            break;
+        }
+    }
+    uint _50 = uint(rayQuery.get_committed_intersection_type());
+    return _50;
+}
+
+fragment void main0(, raytracing::acceleration_structure<raytracing::instancing> topLevelAS [[buffer(0)]])
+{
+    raytracing::intersection_query<raytracing::instancing, raytracing::triangle_data> rayQuery;
+    initFn(rayQuery, topLevelAS);
+    uint _55 = proceeFn(rayQuery);
+}
+
diff --git a/shaders-msl-no-opt/frag/ray-query-mutability.spv14.vk.msl24.frag b/shaders-msl-no-opt/frag/ray-query-mutability.spv14.vk.msl24.frag
new file mode 100644
index 000000000..d35942258
--- /dev/null
+++ b/shaders-msl-no-opt/frag/ray-query-mutability.spv14.vk.msl24.frag
@@ -0,0 +1,26 @@
+#version 460
+
+#extension GL_ARB_separate_shader_objects : enable
+#extension GL_EXT_ray_query : enable
+
+layout(binding = 0) uniform accelerationStructureEXT topLevelAS;
+
+void initFn(rayQueryEXT rayQuery) {
+  vec3  rayOrigin    = vec3(0, 0, 1);
+  vec3  rayDirection = vec3(0, 0,-1);
+  float rayDistance  = 2.0;
+  rayQueryInitializeEXT(rayQuery, topLevelAS, gl_RayFlagsTerminateOnFirstHitEXT, 0xFF, rayOrigin, 0.001, rayDirection, rayDistance);
+}
+
+uint proceeFn(rayQueryEXT rayQuery) {
+  while(rayQueryProceedEXT(rayQuery))
+    ;
+  return rayQueryGetIntersectionTypeEXT(rayQuery, true);
+}
+
+void main() {
+  rayQueryEXT rayQuery;
+
+  initFn(rayQuery);
+  proceeFn(rayQuery);
+}
diff --git a/spirv_msl.cpp b/spirv_msl.cpp
index 8471e34f3..e80f0ef34 100644
--- a/spirv_msl.cpp
+++ b/spirv_msl.cpp
@@ -9366,6 +9366,7 @@ void CompilerMSL::emit_instruction(const Instruction &instruction)
 	case OpRayQueryInitializeKHR:
 	{
 		flush_variable_declaration(ops[0]);
+		register_write(ops[0]);
 		add_spv_func_and_recompile(SPVFuncImplRayQueryIntersectionParams);
 
 		statement(to_expression(ops[0]), ".reset(", "ray(", to_expression(ops[4]), ", ", to_expression(ops[6]), ", ",
@@ -9376,6 +9377,7 @@ void CompilerMSL::emit_instruction(const Instruction &instruction)
 	case OpRayQueryProceedKHR:
 	{
 		flush_variable_declaration(ops[0]);
+		register_write(ops[2]);
 		emit_op(ops[0], ops[1], join(to_expression(ops[2]), ".next()"), false);
 		break;
 	}
@@ -9436,14 +9438,17 @@ void CompilerMSL::emit_instruction(const Instruction &instruction)
 	}
 	case OpRayQueryConfirmIntersectionKHR:
 		flush_variable_declaration(ops[0]);
+		register_write(ops[0]);
 		statement(to_expression(ops[0]), ".commit_triangle_intersection();");
 		break;
 	case OpRayQueryGenerateIntersectionKHR:
 		flush_variable_declaration(ops[0]);
+		register_write(ops[0]);
 		statement(to_expression(ops[0]), ".commit_bounding_box_intersection(", to_expression(ops[1]), ");");
 		break;
 	case OpRayQueryTerminateKHR:
 		flush_variable_declaration(ops[0]);
+		register_write(ops[0]);
 		statement(to_expression(ops[0]), ".abort();");
 		break;
 #undef MSL_RAY_QUERY_GET_OP