Dummy support for RT in metal, add AS encoder

blade-graphics/Cargo.toml

@@ -22,8 +22,7 @@ raw-window-handle = "0.5"
 [target.'cfg(any(target_os = "ios", target_os = "macos"))'.dependencies]
 block = "0.1"
 core-graphics-types = "0.1"
-foreign-types = "0.3"
-metal = "0.24"
+metal = { git = "https://github.com/kvark/metal-rs", branch = "rt" }
 objc = "0.2.5"
 naga = { workspace = true, features = ["msl-out"] }
 

blade-graphics/src/gles/command.rs

@@ -84,6 +84,10 @@ impl super::CommandEncoder {
         }
     }
 
+    pub fn acceleration_structure(&mut self) -> super::PassEncoder<()> {
+        unimplemented!()
+    }
+
     pub fn compute(&mut self) -> super::PassEncoder<super::ComputePipeline> {
         super::PassEncoder {
             commands: &mut self.commands,
@@ -293,6 +297,27 @@ impl crate::traits::TransferEncoder for super::PassEncoder<'_, ()> {
     }
 }
 
+impl super::PassEncoder<'_, ()> {
+    pub fn build_bottom_level(
+        &mut self,
+        _acceleration_structure: super::AccelerationStructure,
+        _meshes: &[crate::AccelerationStructureMesh],
+        _scratch_data: crate::BufferPiece,
+    ) {
+        unimplemented!()
+    }
+
+    pub fn build_top_level(
+        &mut self,
+        _acceleration_structure: super::AccelerationStructure,
+        _instance_count: u32,
+        _instance_data: crate::BufferPiece,
+        _scratch_data: crate::BufferPiece,
+    ) {
+        unimplemented!()
+    }
+}
+
 #[hidden_trait::expose]
 impl crate::traits::PipelineEncoder for super::PipelineEncoder<'_> {
     fn bind<D: crate::ShaderData>(&mut self, group: u32, data: &D) {

blade-graphics/src/metal/command.rs

@@ -62,6 +62,21 @@ impl crate::ShaderBindable for crate::BufferPiece {
         }
     }
 }
+impl crate::ShaderBindable for crate::AccelerationStructure {
+    fn bind_to(&self, ctx: &mut super::PipelineContext, index: u32) {
+        let slot = ctx.targets[index as usize] as _;
+        let value = Some(self.as_ref());
+        if let Some(encoder) = ctx.vs_encoder {
+            encoder.set_vertex_acceleration_structure(slot, value);
+        }
+        if let Some(encoder) = ctx.fs_encoder {
+            encoder.set_fragment_acceleration_structure(slot, value);
+        }
+        if let Some(encoder) = ctx.cs_encoder {
+            encoder.set_acceleration_structure(slot, value);
+        }
+    }
+}
 
 impl super::CommandEncoder {
     pub fn start(&mut self) {
@@ -95,6 +110,20 @@ impl super::CommandEncoder {
         }
     }
 
+    pub fn acceleration_structure(&mut self) -> super::AccelerationStructureCommandEncoder {
+        let raw = objc::rc::autoreleasepool(|| {
+            self.raw
+                .as_mut()
+                .unwrap()
+                .new_acceleration_structure_command_encoder()
+                .to_owned()
+        });
+        super::AccelerationStructureCommandEncoder {
+            raw,
+            phantom: PhantomData,
+        }
+    }
+
     pub fn compute(&mut self) -> super::ComputeCommandEncoder {
         let raw = objc::rc::autoreleasepool(|| {
             self.raw
@@ -272,6 +301,34 @@ impl Drop for super::TransferCommandEncoder<'_> {
     }
 }
 
+impl<'a> super::AccelerationStructureCommandEncoder<'a> {
+    //TODO: move into the trait
+    pub fn build_bottom_level(
+        &mut self,
+        _acceleration_structure: super::AccelerationStructure,
+        _meshes: &[crate::AccelerationStructureMesh],
+        _scratch_data: crate::BufferPiece,
+    ) {
+        unimplemented!()
+    }
+
+    pub fn build_top_level(
+        &mut self,
+        _acceleration_structure: super::AccelerationStructure,
+        _instance_count: u32,
+        _instance_data: crate::BufferPiece,
+        _scratch_data: crate::BufferPiece,
+    ) {
+        unimplemented!()
+    }
+}
+
+impl Drop for super::AccelerationStructureCommandEncoder<'_> {
+    fn drop(&mut self) {
+        self.raw.end_encoding();
+    }
+}
+
 impl super::ComputeCommandEncoder<'_> {
     pub fn with<'p>(
         &'p mut self,

blade-graphics/src/metal/mod.rs

@@ -5,7 +5,7 @@ use std::{
     thread, time,
 };
 
-use foreign_types::ForeignTypeRef as _;
+use metal::foreign_types::{ForeignType as _, ForeignTypeRef as _};
 
 mod command;
 mod pipeline;
@@ -124,6 +124,25 @@ impl Sampler {
     }
 }
 
+#[derive(Clone, Copy, Debug, Hash, PartialEq)]
+pub struct AccelerationStructure {
+    raw: *mut metal::MTLAccelerationStructure,
+}
+
+impl Default for AccelerationStructure {
+    fn default() -> Self {
+        Self {
+            raw: ptr::null_mut(),
+        }
+    }
+}
+
+impl AccelerationStructure {
+    fn as_ref(&self) -> &metal::AccelerationStructureRef {
+        unsafe { metal::AccelerationStructureRef::from_ptr(self.raw) }
+    }
+}
+
 #[derive(Clone, Debug)]
 pub struct SyncPoint {
     cmd_buf: metal::CommandBuffer,
@@ -191,6 +210,12 @@ pub struct TransferCommandEncoder<'a> {
     phantom: PhantomData<&'a CommandEncoder>,
 }
 
+#[derive(Debug)]
+pub struct AccelerationStructureCommandEncoder<'a> {
+    raw: metal::AccelerationStructureCommandEncoder,
+    phantom: PhantomData<&'a CommandEncoder>,
+}
+
 #[derive(Debug)]
 pub struct ComputeCommandEncoder<'a> {
     raw: metal::ComputeCommandEncoder,
@@ -232,6 +257,7 @@ fn map_texture_format(format: crate::TextureFormat) -> metal::MTLPixelFormat {
         Tf::Rgba8Unorm => RGBA8Unorm,
         Tf::Rgba8UnormSrgb => RGBA8Unorm_sRGB,
         Tf::Bgra8UnormSrgb => BGRA8Unorm_sRGB,
+        Tf::Rgba16Float => RGBA16Float,
         Tf::Depth32Float => Depth32Float,
     }
 }

blade-graphics/src/metal/pipeline.rs

@@ -126,6 +126,7 @@ impl super::PipelineLayout {
         let mut num_textures = 0u32;
         let mut num_samplers = 0u32;
         let mut num_buffers = 0u32;
+        let mut num_acceleration_structures = 0u32;
         for layout in bind_group_layouts.iter() {
             let mut targets = Vec::with_capacity(layout.bindings.len());
             for &(_, ref binding) in layout.bindings.iter() {
@@ -143,6 +144,10 @@ impl super::PipelineLayout {
                         num_buffers += 1;
                         num_buffers - 1
                     }
+                    crate::ShaderBinding::AccelerationStructure => {
+                        num_acceleration_structures += 1;
+                        num_acceleration_structures - 1
+                    }
                     crate::ShaderBinding::Plain { .. } => {
                         num_buffers += 1;
                         num_buffers - 1

blade-graphics/src/metal/resource.rs

@@ -64,6 +64,43 @@ fn map_border_color(color: crate::TextureColor) -> metal::MTLSamplerBorderColor
     }
 }
 
+impl super::Context {
+    pub fn get_bottom_level_acceleration_structure_sizes(
+        &self,
+        _meshes: &[crate::AccelerationStructureMesh],
+    ) -> crate::AccelerationStructureSizes {
+        unimplemented!()
+    }
+
+    pub fn get_top_level_acceleration_structure_sizes(
+        &self,
+        _instance_count: u32,
+    ) -> crate::AccelerationStructureSizes {
+        unimplemented!()
+    }
+
+    pub fn create_acceleration_structure_instance_buffer(
+        &self,
+        _instances: &[crate::AccelerationStructureInstance],
+    ) -> super::Buffer {
+        unimplemented!()
+    }
+
+    pub fn create_acceleration_structure(
+        &self,
+        _desc: crate::AccelerationStructureDesc,
+    ) -> super::AccelerationStructure {
+        unimplemented!()
+    }
+
+    pub fn destroy_acceleration_structure(
+        &self,
+        _acceleration_structure: super::AccelerationStructure,
+    ) {
+        unimplemented!()
+    }
+}
+
 #[hidden_trait::expose]
 impl crate::traits::ResourceDevice for super::Context {
     type Buffer = super::Buffer;

blade-graphics/src/vulkan/command.rs

@@ -253,6 +253,14 @@ impl super::CommandEncoder {
         }
     }
 
+    pub fn acceleration_structure(&mut self) -> super::AccelerationStructureCommandEncoder {
+        self.barrier();
+        super::AccelerationStructureCommandEncoder {
+            cmd_buf: self.buffers[0],
+            device: &self.device,
+        }
+    }
+
     pub fn compute(&mut self) -> super::ComputeCommandEncoder {
         self.barrier();
         super::ComputeCommandEncoder {
@@ -418,23 +426,9 @@ impl crate::traits::TransferEncoder for super::TransferCommandEncoder<'_> {
     }
 }
 
-impl<'a> super::ComputeCommandEncoder<'a> {
-    pub fn with<'b, 'p>(
-        &'b mut self,
-        pipeline: &'p super::ComputePipeline,
-    ) -> super::PipelineEncoder<'b, 'p> {
-        super::PipelineEncoder {
-            cmd_buf: self.cmd_buf,
-            layout: &pipeline.layout,
-            bind_point: vk::PipelineBindPoint::COMPUTE,
-            device: self.device,
-            update_data: self.update_data,
-        }
-        .init(pipeline.raw)
-    }
-
+impl<'a> super::AccelerationStructureCommandEncoder<'a> {
     //TODO: move into the trait
-    pub fn build_bottom_level_acceleration_structure(
+    pub fn build_bottom_level(
         &mut self,
         acceleration_structure: super::AccelerationStructure,
         meshes: &[crate::AccelerationStructureMesh],
@@ -456,7 +450,7 @@ impl<'a> super::ComputeCommandEncoder<'a> {
         }
     }
 
-    pub fn build_top_level_acceleration_structure(
+    pub fn build_top_level(
         &mut self,
         acceleration_structure: super::AccelerationStructure,
         instance_count: u32,
@@ -501,6 +495,22 @@ impl<'a> super::ComputeCommandEncoder<'a> {
     }
 }
 
+impl<'a> super::ComputeCommandEncoder<'a> {
+    pub fn with<'b, 'p>(
+        &'b mut self,
+        pipeline: &'p super::ComputePipeline,
+    ) -> super::PipelineEncoder<'b, 'p> {
+        super::PipelineEncoder {
+            cmd_buf: self.cmd_buf,
+            layout: &pipeline.layout,
+            bind_point: vk::PipelineBindPoint::COMPUTE,
+            device: self.device,
+            update_data: self.update_data,
+        }
+        .init(pipeline.raw)
+    }
+}
+
 impl<'a> super::RenderCommandEncoder<'a> {
     pub fn set_scissor_rect(&mut self, rect: &crate::ScissorRect) {
         let vk_scissor = vk::Rect2D {

blade-graphics/src/vulkan/mod.rs

@@ -213,6 +213,10 @@ pub struct TransferCommandEncoder<'a> {
     raw: vk::CommandBuffer,
     device: &'a Device,
 }
+pub struct AccelerationStructureCommandEncoder<'a> {
+    raw: vk::CommandBuffer,
+    device: &'a Device,
+}
 pub struct ComputeCommandEncoder<'a> {
     cmd_buf: CommandBuffer,
     device: &'a Device,

examples/ray-trace/main.rs

@@ -188,12 +188,12 @@ impl Example {
             buffer_count: 2,
         });
         command_encoder.start();
-        if let mut pass = command_encoder.compute() {
-            pass.build_bottom_level_acceleration_structure(blas, &meshes, scratch_buffer.at(0));
+        if let mut pass = command_encoder.acceleration_structure() {
+            pass.build_bottom_level(blas, &meshes, scratch_buffer.at(0));
         }
         //Note: separate pass in order to enforce synchronization
-        if let mut pass = command_encoder.compute() {
-            pass.build_top_level_acceleration_structure(
+        if let mut pass = command_encoder.acceleration_structure() {
+            pass.build_top_level(
                 tlas,
                 instances.len() as u32,
                 instance_buffer.at(0),

tests/parse_shaders.rs

@@ -41,7 +41,7 @@ fn parse_wgsl() {
             //TODO: re-use the validator
             Validator::new(
                 naga::valid::ValidationFlags::all() ^ naga::valid::ValidationFlags::BINDINGS,
-                naga::valid::Capabilities::empty(),
+                naga::valid::Capabilities::RAY_QUERY,
             )
             .validate(&module)
             .unwrap_or_else(|e| {