From 1ecf79a2bc9c938a3a13c5e68f976e580664663d Mon Sep 17 00:00:00 2001 From: Dzmitry Malyshau Date: Sat, 24 Jan 2026 21:46:24 -0800 Subject: [PATCH 1/2] Support TLAS binding arrays --- Cargo.toml | 2 +- blade-graphics/src/derive.rs | 3 +++ blade-graphics/src/gles/command.rs | 7 +++++++ blade-graphics/src/gles/pipeline.rs | 8 ++++++-- blade-graphics/src/lib.rs | 3 +++ blade-graphics/src/metal/command.rs | 7 +++++++ blade-graphics/src/metal/pipeline.rs | 6 ++++-- blade-graphics/src/shader.rs | 10 +++++++++- blade-graphics/src/vulkan/command.rs | 15 +++++++++++++++ blade-graphics/src/vulkan/pipeline.rs | 10 +++++++++- blade-render/code/fill-gbuf.wgsl | 1 + blade-render/code/ray-trace.wgsl | 1 + examples/ray-query/shader.wgsl | 1 + 13 files changed, 67 insertions(+), 7 deletions(-) diff --git a/Cargo.toml b/Cargo.toml index ed83a5dc..b57f1560 100644 --- a/Cargo.toml +++ b/Cargo.toml @@ -21,7 +21,7 @@ glam = { version = "0.30", features = ["mint"] } gltf = { version = "1.1", default-features = false } log = "0.4" mint = "0.5" -naga = { version = "27.0", features = ["wgsl-in", "termcolor"] } +naga = { git = "https://github.com/gfx-rs/wgpu", rev = "760032a20d6d742d1df28136f36a673ab2584929", features = ["wgsl-in", "termcolor"] } profiling = "1" slab = "0.4" strum = { version = "0.27", features = ["derive"] } diff --git a/blade-graphics/src/derive.rs b/blade-graphics/src/derive.rs index 6fabfb31..663935c1 100644 --- a/blade-graphics/src/derive.rs +++ b/blade-graphics/src/derive.rs @@ -28,6 +28,9 @@ impl<'a, const N: ResourceIndex> HasShaderBinding for &'a super::TextureArray impl HasShaderBinding for super::AccelerationStructure { const TYPE: ShaderBinding = ShaderBinding::AccelerationStructure; } +impl<'a, const N: ResourceIndex> HasShaderBinding for &'a super::AccelerationStructureArray { + const TYPE: ShaderBinding = ShaderBinding::AccelerationStructureArray { count: N }; +} pub trait HasVertexAttribute { const FORMAT: VertexFormat; diff --git a/blade-graphics/src/gles/command.rs b/blade-graphics/src/gles/command.rs index 783f3bff..fad01bce 100644 --- a/blade-graphics/src/gles/command.rs +++ b/blade-graphics/src/gles/command.rs @@ -79,6 +79,13 @@ impl crate::ShaderBindable for super::AccelerationStructure { } } } +impl<'a, const N: crate::ResourceIndex> crate::ShaderBindable + for &'a crate::AccelerationStructureArray +{ + fn bind_to(&self, _ctx: &mut super::PipelineContext, _index: u32) { + unimplemented!() + } +} impl super::CommandEncoder { fn begin_pass(&mut self, label: &str) { diff --git a/blade-graphics/src/gles/pipeline.rs b/blade-graphics/src/gles/pipeline.rs index 76bd4344..d1f63bcb 100644 --- a/blade-graphics/src/gles/pipeline.rs +++ b/blade-graphics/src/gles/pipeline.rs @@ -76,7 +76,10 @@ impl super::Context { } crate::ShaderBinding::TextureArray { .. } | crate::ShaderBinding::BufferArray { .. } - | crate::ShaderBinding::AccelerationStructure => unimplemented!(), + | crate::ShaderBinding::AccelerationStructure + | crate::ShaderBinding::AccelerationStructureArray { .. } => { + unimplemented!() + } crate::ShaderBinding::Plain { .. } => { num_buffers += 1; num_buffers - 1 @@ -248,7 +251,8 @@ impl super::Context { } crate::ShaderBinding::TextureArray { .. } | crate::ShaderBinding::BufferArray { .. } - | crate::ShaderBinding::AccelerationStructure => { + | crate::ShaderBinding::AccelerationStructure + | crate::ShaderBinding::AccelerationStructureArray { .. } => { unimplemented!() } crate::ShaderBinding::Plain { size } => { diff --git a/blade-graphics/src/lib.rs b/blade-graphics/src/lib.rs index 08dc41a2..405e0ca0 100644 --- a/blade-graphics/src/lib.rs +++ b/blade-graphics/src/lib.rs @@ -281,6 +281,8 @@ impl std::ops::IndexMut for ResourceAr } pub type BufferArray = ResourceArray; pub type TextureArray = ResourceArray; +pub type AccelerationStructureArray = + ResourceArray; #[derive(Clone, Copy, Debug)] pub struct TexturePiece { @@ -644,6 +646,7 @@ pub enum ShaderBinding { Buffer, BufferArray { count: u32 }, AccelerationStructure, + AccelerationStructureArray { count: u32 }, Plain { size: u32 }, } diff --git a/blade-graphics/src/metal/command.rs b/blade-graphics/src/metal/command.rs index 43ebc630..99321b42 100644 --- a/blade-graphics/src/metal/command.rs +++ b/blade-graphics/src/metal/command.rs @@ -102,6 +102,13 @@ impl crate::ShaderBindable for crate::AccelerationStructure { } } } +impl<'a, const N: crate::ResourceIndex> crate::ShaderBindable + for &'a crate::AccelerationStructureArray +{ + fn bind_to(&self, _ctx: &mut super::PipelineContext, _index: u32) { + unimplemented!() + } +} impl super::TimingData { fn add(&mut self, label: &str) -> usize { diff --git a/blade-graphics/src/metal/pipeline.rs b/blade-graphics/src/metal/pipeline.rs index 2b59f688..e283cfc2 100644 --- a/blade-graphics/src/metal/pipeline.rs +++ b/blade-graphics/src/metal/pipeline.rs @@ -157,7 +157,8 @@ fn make_pipeline_layout( num_buffers - 1 } crate::ShaderBinding::TextureArray { .. } - | crate::ShaderBinding::BufferArray { .. } => unimplemented!(), + | crate::ShaderBinding::BufferArray { .. } + | crate::ShaderBinding::AccelerationStructureArray { .. } => unimplemented!(), crate::ShaderBinding::AccelerationStructure => { num_buffers += 1; num_buffers - 1 @@ -268,7 +269,8 @@ impl super::Context { ..Default::default() }, crate::ShaderBinding::TextureArray { .. } - | crate::ShaderBinding::BufferArray { .. } => todo!(), + | crate::ShaderBinding::BufferArray { .. } + | crate::ShaderBinding::AccelerationStructureArray { .. } => todo!(), }; naga_resources.resources.insert(res_binding, bind_target); } diff --git a/blade-graphics/src/shader.rs b/blade-graphics/src/shader.rs index fc5cecbc..fea93da2 100644 --- a/blade-graphics/src/shader.rs +++ b/blade-graphics/src/shader.rs @@ -28,7 +28,9 @@ impl super::Context { let flags = naga::valid::ValidationFlags::all() ^ naga::valid::ValidationFlags::BINDINGS; let mut caps = naga::valid::Capabilities::empty(); caps.set( - naga::valid::Capabilities::RAY_QUERY | naga::valid::Capabilities::SAMPLED_TEXTURE_AND_STORAGE_BUFFER_ARRAY_NON_UNIFORM_INDEXING, + naga::valid::Capabilities::RAY_QUERY + | naga::valid::Capabilities::TEXTURE_AND_SAMPLER_BINDING_ARRAY_NON_UNIFORM_INDEXING + | naga::valid::Capabilities::STORAGE_BUFFER_BINDING_ARRAY_NON_UNIFORM_INDEXING, !device_caps.ray_query.is_empty(), ); caps.set( @@ -170,6 +172,7 @@ impl super::Shader { let count = match proto_binding { crate::ShaderBinding::TextureArray { count } => count, crate::ShaderBinding::BufferArray { count } => count, + crate::ShaderBinding::AccelerationStructureArray { count } => count, _ => 0, }; let proto = match module.types[base].inner { @@ -179,6 +182,9 @@ impl super::Shader { naga::TypeInner::Struct { .. } => { crate::ShaderBinding::BufferArray { count } } + naga::TypeInner::AccelerationStructure { .. } => { + crate::ShaderBinding::AccelerationStructureArray { count } + } ref other => panic!("Unsupported binding array for {:?}", other), }; (proto, var_access) @@ -275,6 +281,7 @@ impl super::Shader { interpolation: None, sampling: None, blend_src: None, + per_primitive: false, }; for (buffer_index, vertex_fetch) in fetch_states.iter().enumerate() { for (attribute_index, &(at_name, _)) in @@ -307,6 +314,7 @@ impl super::Shader { interpolation: None, sampling: None, blend_src: None, + per_primitive: false, }); location += 1; } diff --git a/blade-graphics/src/vulkan/command.rs b/blade-graphics/src/vulkan/command.rs index 8ac451d1..e5deba35 100644 --- a/blade-graphics/src/vulkan/command.rs +++ b/blade-graphics/src/vulkan/command.rs @@ -130,6 +130,21 @@ impl crate::ShaderBindable for super::AccelerationStructure { ctx.write(index, self.raw); } } +impl<'a, const N: crate::ResourceIndex> crate::ShaderBindable + for &'a crate::AccelerationStructureArray +{ + fn bind_to(&self, ctx: &mut super::PipelineContext, index: u32) { + assert!(self.data.len() <= N as usize); + ctx.write_array( + index, + self.data + .iter() + .map(|accel| accel.raw) + .cycle() + .take(N as usize), + ); + } +} impl crate::TexturePiece { fn subresource_layers(&self) -> vk::ImageSubresourceLayers { diff --git a/blade-graphics/src/vulkan/pipeline.rs b/blade-graphics/src/vulkan/pipeline.rs index 76a11285..e3fe686d 100644 --- a/blade-graphics/src/vulkan/pipeline.rs +++ b/blade-graphics/src/vulkan/pipeline.rs @@ -20,7 +20,8 @@ impl super::Context { for (binding_index, &(_, binding)) in layout.bindings.iter().enumerate() { let binding_array_size = match binding { crate::ShaderBinding::TextureArray { count } - | crate::ShaderBinding::BufferArray { count } => Some(count), + | crate::ShaderBinding::BufferArray { count } + | crate::ShaderBinding::AccelerationStructureArray { count } => Some(count), _ => None, }; let rb = naga::ResourceBinding { @@ -51,6 +52,7 @@ impl super::Context { bounds_check_policies: naga::proc::BoundsCheckPolicies::default(), zero_initialize_workgroup_memory: spv::ZeroInitializeWorkgroupMemoryMode::None, force_loop_bounding: false, + ray_query_initialization_tracking: true, use_storage_input_output_16: false, debug_info: None, } @@ -241,6 +243,12 @@ impl super::Context { 1u32, vk::DescriptorBindingFlags::empty(), ), + crate::ShaderBinding::AccelerationStructureArray { count } => ( + vk::DescriptorType::ACCELERATION_STRUCTURE_KHR, + mem::size_of::(), + count, + vk::DescriptorBindingFlags::PARTIALLY_BOUND, + ), crate::ShaderBinding::Plain { size } => ( vk::DescriptorType::INLINE_UNIFORM_BLOCK_EXT, 1, diff --git a/blade-render/code/fill-gbuf.wgsl b/blade-render/code/fill-gbuf.wgsl index 37b2a83b..7d400361 100644 --- a/blade-render/code/fill-gbuf.wgsl +++ b/blade-render/code/fill-gbuf.wgsl @@ -1,3 +1,4 @@ +enable wgpu_ray_query; #include "quaternion.inc.wgsl" #include "camera.inc.wgsl" #include "debug.inc.wgsl" diff --git a/blade-render/code/ray-trace.wgsl b/blade-render/code/ray-trace.wgsl index 94e3e21a..9bd3fbf0 100644 --- a/blade-render/code/ray-trace.wgsl +++ b/blade-render/code/ray-trace.wgsl @@ -1,3 +1,4 @@ +enable wgpu_ray_query; #include "quaternion.inc.wgsl" #include "random.inc.wgsl" #include "env-importance.inc.wgsl" diff --git a/examples/ray-query/shader.wgsl b/examples/ray-query/shader.wgsl index 8387ab08..11f44593 100644 --- a/examples/ray-query/shader.wgsl +++ b/examples/ray-query/shader.wgsl @@ -1,3 +1,4 @@ +enable wgpu_ray_query; const MAX_BOUNCES: i32 = 3; struct Parameters { From 9a1afac64c04635886a62a7ff746de969700bf3b Mon Sep 17 00:00:00 2001 From: Dzmitry Malyshau Date: Sun, 25 Jan 2026 15:49:58 -0800 Subject: [PATCH 2/2] Separate Capabilities flag for binding arrays --- blade-graphics/src/gles/mod.rs | 1 + blade-graphics/src/lib.rs | 4 +++- blade-graphics/src/metal/mod.rs | 1 + blade-graphics/src/shader.rs | 8 +++++++- blade-graphics/src/vulkan/init.rs | 24 +++++++++++++++++++++--- blade-graphics/src/vulkan/mod.rs | 1 + docs/CHANGELOG.md | 4 ++++ 7 files changed, 38 insertions(+), 5 deletions(-) diff --git a/blade-graphics/src/gles/mod.rs b/blade-graphics/src/gles/mod.rs index aa97790f..7cdb6bde 100644 --- a/blade-graphics/src/gles/mod.rs +++ b/blade-graphics/src/gles/mod.rs @@ -444,6 +444,7 @@ struct ExecutionContext { impl Context { pub fn capabilities(&self) -> crate::Capabilities { crate::Capabilities { + binding_array: false, ray_query: crate::ShaderVisibility::empty(), sample_count_mask: 0x1 | 0x4, //TODO: accurate info dual_source_blending: false, diff --git a/blade-graphics/src/lib.rs b/blade-graphics/src/lib.rs index 405e0ca0..71b4568b 100644 --- a/blade-graphics/src/lib.rs +++ b/blade-graphics/src/lib.rs @@ -122,7 +122,9 @@ impl From for NotSupportedError { #[derive(Clone, Debug, Default, PartialEq)] pub struct Capabilities { - /// Which shader stages support ray queries + /// Support binding arrays of handles. + pub binding_array: bool, + /// Which shader stages support ray queries. pub ray_query: ShaderVisibility, /// Bit mask of supported MSAA sample counts. pub sample_count_mask: u32, diff --git a/blade-graphics/src/metal/mod.rs b/blade-graphics/src/metal/mod.rs index 200ba948..e3a0ddec 100644 --- a/blade-graphics/src/metal/mod.rs +++ b/blade-graphics/src/metal/mod.rs @@ -519,6 +519,7 @@ impl Context { let device = self.device.lock().unwrap(); crate::Capabilities { + binding_array: false, ray_query: if device.supportsFamily(metal::MTLGPUFamily::Apple6) { crate::ShaderVisibility::all() } else if device.supportsFamily(metal::MTLGPUFamily::Mac2) diff --git a/blade-graphics/src/shader.rs b/blade-graphics/src/shader.rs index fea93da2..29549eb9 100644 --- a/blade-graphics/src/shader.rs +++ b/blade-graphics/src/shader.rs @@ -28,9 +28,15 @@ impl super::Context { let flags = naga::valid::ValidationFlags::all() ^ naga::valid::ValidationFlags::BINDINGS; let mut caps = naga::valid::Capabilities::empty(); caps.set( - naga::valid::Capabilities::RAY_QUERY + naga::valid::Capabilities::STORAGE_BUFFER_BINDING_ARRAY + | naga::valid::Capabilities::TEXTURE_AND_SAMPLER_BINDING_ARRAY | naga::valid::Capabilities::TEXTURE_AND_SAMPLER_BINDING_ARRAY_NON_UNIFORM_INDEXING | naga::valid::Capabilities::STORAGE_BUFFER_BINDING_ARRAY_NON_UNIFORM_INDEXING, + device_caps.binding_array, + ); + caps.set( + naga::valid::Capabilities::RAY_QUERY + | naga::valid::Capabilities::ACCELERATION_STRUCTURE_BINDING_ARRAY, !device_caps.ray_query.is_empty(), ); caps.set( diff --git a/blade-graphics/src/vulkan/init.rs b/blade-graphics/src/vulkan/init.rs index 1fa1665e..fbc5cabd 100644 --- a/blade-graphics/src/vulkan/init.rs +++ b/blade-graphics/src/vulkan/init.rs @@ -44,6 +44,7 @@ struct AdapterCapabilities { device_information: crate::DeviceInformation, queue_family_index: u32, layered: bool, + binding_array: bool, ray_tracing: Option, buffer_marker: bool, shader_info: bool, @@ -209,6 +210,14 @@ unsafe fn inspect_adapter( true }; + let supports_descriptor_indexing = api_version >= vk::API_VERSION_1_2 + || supported_extensions.contains(&vk::EXT_DESCRIPTOR_INDEXING_NAME); + let binding_array = supports_descriptor_indexing + && descriptor_indexing_features.descriptor_binding_partially_bound == vk::TRUE + && descriptor_indexing_features.shader_storage_buffer_array_non_uniform_indexing + == vk::TRUE + && descriptor_indexing_features.shader_sampled_image_array_non_uniform_indexing == vk::TRUE; + let ray_tracing = if !supported_extensions.contains(&vk::KHR_ACCELERATION_STRUCTURE_NAME) || !supported_extensions.contains(&vk::KHR_RAY_QUERY_NAME) { @@ -277,6 +286,7 @@ unsafe fn inspect_adapter( device_information, queue_family_index, layered: portability_subset_properties.min_vertex_input_binding_stride_alignment != 0, + binding_array, ray_tracing, buffer_marker, shader_info, @@ -457,9 +467,13 @@ impl super::Context { log::info!("Enabling Vulkan Portability"); device_extensions.push(vk::KHR_PORTABILITY_SUBSET_NAME); } + let needs_descriptor_indexing = + capabilities.binding_array || capabilities.ray_tracing.is_some(); + if needs_descriptor_indexing && capabilities.api_version < vk::API_VERSION_1_2 { + device_extensions.push(vk::EXT_DESCRIPTOR_INDEXING_NAME); + } if capabilities.ray_tracing.is_some() { if capabilities.api_version < vk::API_VERSION_1_2 { - device_extensions.push(vk::EXT_DESCRIPTOR_INDEXING_NAME); device_extensions.push(vk::KHR_BUFFER_DEVICE_ADDRESS_NAME); device_extensions.push(vk::KHR_SHADER_FLOAT_CONTROLS_NAME); device_extensions.push(vk::KHR_SPIRV_1_4_NAME); @@ -514,13 +528,16 @@ impl super::Context { let mut khr_buffer_device_address; let mut khr_acceleration_structure; let mut khr_ray_query; - if capabilities.ray_tracing.is_some() { + if needs_descriptor_indexing { ext_descriptor_indexing = vk::PhysicalDeviceDescriptorIndexingFeaturesEXT { shader_storage_buffer_array_non_uniform_indexing: vk::TRUE, shader_sampled_image_array_non_uniform_indexing: vk::TRUE, descriptor_binding_partially_bound: vk::TRUE, ..Default::default() }; + device_create_info = device_create_info.push_next(&mut ext_descriptor_indexing); + } + if capabilities.ray_tracing.is_some() { khr_buffer_device_address = vk::PhysicalDeviceBufferDeviceAddressFeaturesKHR { buffer_device_address: vk::TRUE, ..Default::default() @@ -534,7 +551,6 @@ impl super::Context { ..Default::default() }; device_create_info = device_create_info - .push_next(&mut ext_descriptor_indexing) .push_next(&mut khr_buffer_device_address) .push_next(&mut khr_acceleration_structure) .push_next(&mut khr_ray_query); @@ -747,6 +763,7 @@ impl super::Context { .limits .framebuffer_depth_sample_counts, dual_source_blending: capabilities.dual_source_blending, + binding_array: capabilities.binding_array, instance, entry, }) @@ -766,6 +783,7 @@ impl super::Context { pub fn capabilities(&self) -> crate::Capabilities { crate::Capabilities { + binding_array: self.binding_array, ray_query: match self.device.ray_tracing { Some(_) => crate::ShaderVisibility::all(), None => crate::ShaderVisibility::empty(), diff --git a/blade-graphics/src/vulkan/mod.rs b/blade-graphics/src/vulkan/mod.rs index 5ce50eb4..808e8ff0 100644 --- a/blade-graphics/src/vulkan/mod.rs +++ b/blade-graphics/src/vulkan/mod.rs @@ -159,6 +159,7 @@ pub struct Context { min_buffer_alignment: u64, sample_count_flags: vk::SampleCountFlags, dual_source_blending: bool, + binding_array: bool, instance: Instance, entry: ash::Entry, } diff --git a/docs/CHANGELOG.md b/docs/CHANGELOG.md index 103ba5c6..30d9e611 100644 --- a/docs/CHANGELOG.md +++ b/docs/CHANGELOG.md @@ -1,5 +1,9 @@ Changelog for *Blade* project +## blade-graphics-0.8 (TBD) +- graphics + - separate `Capabilities` flag for binding arrays, including TLAS arrays + ## blade-graphics-0.7 (27 Sep 2025) - graphics