radv/rt: Don't consider non-internal INTERSECTION shaders as the traversal shader

Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/39314>

src/amd/vulkan/nir/radv_nir_rt_traversal_shader.c

@@ -1145,7 +1145,6 @@ radv_build_traversal_shader(struct radv_device *device, struct radv_ray_tracing_
    /* Create the traversal shader as an intersection shader to prevent validation failures due to
     * invalid variable modes.*/
    nir_builder b = radv_meta_nir_init_shader(device, MESA_SHADER_INTERSECTION, "rt_traversal");
-   b.shader->info.internal = false;
    b.shader->info.workgroup_size[0] = pdev->rt_wave_size;
    b.shader->info.api_subgroup_size = pdev->rt_wave_size;
    b.shader->info.max_subgroup_size = pdev->rt_wave_size;

src/amd/vulkan/radv_pipeline.c

@@ -762,7 +762,10 @@ radv_GetPipelineExecutablePropertiesKHR(VkDevice _device, const VkPipelineInfoKH
          description = "Vulkan Miss Shader";
          break;
       case MESA_SHADER_INTERSECTION:
-         description = "Shader responsible for traversing the acceleration structure";
+         if (shader->info.type == RADV_SHADER_TYPE_RT_TRAVERSAL)
+            description = "Shader responsible for traversing the acceleration structure";
+         else
+            description = "Vulkan Intersection Shader";
          break;
       case MESA_SHADER_CALLABLE:
          description = "Vulkan Callable Shader";

src/amd/vulkan/radv_pipeline_rt.c

@@ -400,6 +400,7 @@ radv_rt_nir_to_asm(struct radv_device *device, struct vk_pipeline_cache *cache,
 
    stage->info.user_sgprs_locs = stage->args.user_sgprs_locs;
    stage->info.inline_push_constant_mask = stage->args.ac.inline_push_const_mask;
+   stage->info.type = radv_is_traversal_shader(stage->nir) ? RADV_SHADER_TYPE_RT_TRAVERSAL : RADV_SHADER_TYPE_DEFAULT;
 
    /* Move ray tracing system values to the top that are set by rt_trace_ray
     * to prevent them from being overwritten by other rt_trace_ray calls.
@@ -469,7 +470,7 @@ radv_rt_nir_to_asm(struct radv_device *device, struct vk_pipeline_cache *cache,
       NIR_PASS(_, stage->nir, nir_lower_vars_to_ssa);
       NIR_PASS(_, stage->nir, nir_opt_copy_prop);
       NIR_PASS(_, stage->nir, nir_opt_remove_phis);
-      if (!stage->key.optimisations_disabled)
+      if (!stage->key.optimisations_disabled && !radv_is_traversal_shader(stage->nir))
          NIR_PASS(_, stage->nir, nir_minimize_call_live_states);
 
       stage->info.nir_shared_size = MAX2(stage->info.nir_shared_size, temp_stage.info.nir_shared_size);
@@ -481,7 +482,7 @@ radv_rt_nir_to_asm(struct radv_device *device, struct vk_pipeline_cache *cache,
    bool dump_nir = dump_shader && (instance->debug_flags & RADV_DEBUG_DUMP_NIR);
    bool replayable = (pipeline->base.base.create_flags &
                       VK_PIPELINE_CREATE_2_RAY_TRACING_SHADER_GROUP_HANDLE_CAPTURE_REPLAY_BIT_KHR) &&
-                     stage->stage != MESA_SHADER_INTERSECTION;
+                     !radv_is_traversal_shader(stage->nir);
 
    if (dump_shader) {
       simple_mtx_lock(&instance->shader_dump_mtx);

src/amd/vulkan/radv_shader.c

@@ -106,9 +106,21 @@ vectorize_vec2_16bit(const nir_instr *instr, const void *_)
       return 1;
 }
 
+bool
+radv_is_traversal_shader(nir_shader *nir)
+{
+   return nir && nir->info.stage == MESA_SHADER_INTERSECTION && nir->info.internal;
+}
+
 static bool
 is_meta_shader(nir_shader *nir)
 {
+   /* The built-in traversal shader is marked as "internal", to distinguish
+    * it from intersection shaders even though both share the INTERSECTION
+    * shader stage. It is not a meta shader, though, so special-case it here.
+    */
+   if (radv_is_traversal_shader(nir))
+      return false;
    return nir && nir->info.internal;
 }
 
@@ -141,7 +153,7 @@ radv_can_dump_shader(struct radv_device *device, nir_shader *nir)
    const struct radv_physical_device *pdev = radv_device_physical(device);
    const struct radv_instance *instance = radv_physical_device_instance(pdev);
 
-   if (is_meta_shader(nir))
+   if (is_meta_shader(nir) && nir->info.stage != MESA_SHADER_INTERSECTION)
       return instance->debug_flags & RADV_DEBUG_DUMP_META_SHADERS;
 
    if (!nir)

src/amd/vulkan/radv_shader.h

@@ -583,6 +583,8 @@ unsigned radv_compute_spi_ps_input(const struct radv_physical_device *pdev,
                                    const struct radv_graphics_state_key *gfx_state,
                                    const struct radv_shader_info *info);
 
+bool radv_is_traversal_shader(nir_shader *nir);
+
 bool radv_can_dump_shader(struct radv_device *device, nir_shader *nir);
 
 bool radv_can_dump_shader_stats(struct radv_device *device, nir_shader *nir);

src/amd/vulkan/radv_shader_info.h

@@ -32,6 +32,7 @@ enum radv_shader_type {
    RADV_SHADER_TYPE_GS_COPY,
    RADV_SHADER_TYPE_TRAP_HANDLER,
    RADV_SHADER_TYPE_RT_PROLOG,
+   RADV_SHADER_TYPE_RT_TRAVERSAL,
 };
 
 struct radv_vs_output_info {