WGPU 0.19

Cargo.toml

@@ -17,7 +17,7 @@ bytemuck = "1.7"
 cfg-if = "1.0"
 futures = { version = "0.3", default-features = false, features = ["executor"] }
 image = { version = "0.24.6", default-features = false, optional = true }
-wgpu = { version = "0.14", features = ["spirv"] }
+wgpu = { version = "0.19.1", features = ["spirv"] }
 ndarray = { version = "0.15", default-features = false, features = [
     "std",
 ], optional = true }

examples/image-compatibility/shader.wgsl

@@ -7,7 +7,7 @@ fn main(@builtin(global_invocation_id) global_id: vec3<u32>) {
     let pixel = textureLoad(input, coord, 0);
 
     let dims = textureDimensions(input);
-    let mirror_coord = vec2<i32>(dims.x - coord.x, coord.y);
+    let mirror_coord = vec2<i32>(i32(dims.x) - coord.x, coord.y);
 
     textureStore(output, mirror_coord, pixel);
 }

examples/mirror-image/shader.wgsl

@@ -7,7 +7,7 @@ fn main(@builtin(global_invocation_id) global_id: vec3<u32>) {
     let pixel = textureLoad(input, coord, 0);
 
     let dims = textureDimensions(input);
-    let mirror_coord = vec2<i32>(dims.x - coord.x, coord.y);
+    let mirror_coord = vec2<i32>(i32(dims.x) - coord.x, coord.y);
 
     textureStore(output, mirror_coord, pixel);
 }

examples/webcam/shader.wgsl

@@ -6,15 +6,15 @@ struct Time {
 @group(0) @binding(1) var output: texture_storage_2d<rgba8unorm, write>;
 @group(0) @binding(2) var<uniform> time: Time; 
 
-let pi: f32 = 3.14159;
+const pi: f32 = 3.14159;
 
 @compute @workgroup_size(32, 32, 1)
 fn main(@builtin(global_invocation_id) global_id: vec3<u32>) {
     let coord = vec2<i32>(global_id.xy);
     let pixel = textureLoad(input, coord, 0);
 
     let dims = textureDimensions(input);
-    let mirror_coord = vec2<i32>(dims.x - coord.x, coord.y);
+    let mirror_coord = vec2<i32>(i32(dims.x) - coord.x, coord.y);
 
     let t = pi * time.time;
     let colour = vec4<f32>(sin(t), sin(0.25 * t), sin(0.5 * t), 1.0);

src/features/integrate_image.rs

@@ -112,12 +112,7 @@ where
         Container: std::ops::Deref<Target = [Pixel::Subpixel]>,
     {
         let (width, height) = img.dimensions();
-        GpuConstImage::from_bytes(
-            fw,
-            bytemuck::cast_slice(img),
-            width * Pixel::GpgpuPixel::byte_size() as u32,
-            height,
-        )
+        GpuConstImage::from_bytes(fw, bytemuck::cast_slice(img), width, height)
     }
 
     /// Constructs a new normalised [`GpuConstImage`] from a [`image::ImageBuffer`].
@@ -228,7 +223,7 @@ where
     }
 }
 
-pub(self) fn bytes_to_primitive_vec<P>(mut bytes: Vec<u8>) -> Vec<P::Subpixel>
+fn bytes_to_primitive_vec<P>(mut bytes: Vec<u8>) -> Vec<P::Subpixel>
 where
     P: image::Pixel,
     P::Subpixel: bytemuck::Pod,

src/framework.rs

@@ -4,16 +4,14 @@ use crate::Framework;
 
 impl Default for Framework {
     fn default() -> Self {
-        let backend = wgpu::util::backend_bits_from_env().unwrap_or(wgpu::Backends::PRIMARY);
         let power_preference = wgpu::util::power_preference_from_env()
             .unwrap_or(wgpu::PowerPreference::HighPerformance);
-        let instance = wgpu::Instance::new(backend);
+        let instance = wgpu::Instance::new(wgpu::InstanceDescriptor {
+            backends: wgpu::Backends::PRIMARY,
+            ..Default::default()
+        });
 
-        log::debug!(
-            "Requesting device with {:#?} and {:#?}",
-            backend,
-            power_preference
-        );
+        log::debug!("Requesting device with {:#?}", power_preference);
 
         futures::executor::block_on(async {
             let adapter = instance
@@ -38,8 +36,8 @@ impl Framework {
             .request_device(
                 &wgpu::DeviceDescriptor {
                     label: None,
-                    features: adapter.features(), // Change this to allow proper WebGL2 support (in the future™️).
-                    limits: adapter.limits(),     // Bye WebGL2 support :(
+                    required_features: adapter.features(), // Change this to allow proper WebGL2 support (in the future™️).
+                    required_limits: adapter.limits(),     // Bye WebGL2 support :(
                 },
                 None,
             )

src/kernel.rs

@@ -352,6 +352,7 @@ impl<'fw> Kernel<'fw> {
         {
             let mut cpass = encoder.begin_compute_pass(&wgpu::ComputePassDescriptor {
                 label: Some("Kernel::enqueue"),
+                timestamp_writes: None,
             });
 
             cpass.set_pipeline(&self.pipeline);

src/primitives/buffers.rs

@@ -56,7 +56,7 @@ where
     }
 
     fn from_slice(fw: &'fw crate::Framework, slice: &[T]) -> Self {
-        let size = (slice.len() * std::mem::size_of::<T>()) as u64;
+        let size = std::mem::size_of_val(slice) as u64;
         let buf = fw
             .device
             .create_buffer_init(&wgpu::util::BufferInitDescriptor {
@@ -93,7 +93,7 @@ where
 {
     /// Pulls some elements from the [`GpuBuffer`] into `buf`, returning how many elements were read.
     pub async fn read(&self, buf: &mut [T]) -> BufferResult<u64> {
-        let output_size = (buf.len() * std::mem::size_of::<T>()) as u64;
+        let output_size = std::mem::size_of_val(buf) as u64;
         let download_size = if output_size > self.size {
             self.size
         } else {
@@ -141,7 +141,7 @@ where
     /// This function will attempt to write the entire contents of `buf` unless its capacity
     /// exceeds the one of the source buffer, in which case `GpuBuffer::capacity()` elements are written.
     pub fn write(&self, buf: &[T]) -> BufferResult<u64> {
-        let input_size = (buf.len() * std::mem::size_of::<T>()) as u64;
+        let input_size = std::mem::size_of_val(buf) as u64;
         let upload_size = if input_size > self.size {
             self.size
         } else {
@@ -195,7 +195,7 @@ where
     }
 
     fn from_slice(fw: &'fw crate::Framework, slice: &[T]) -> Self {
-        let size = (slice.len() * std::mem::size_of::<T>()) as u64;
+        let size = std::mem::size_of_val(slice) as u64;
         let buf = fw
             .device
             .create_buffer_init(&wgpu::util::BufferInitDescriptor {
@@ -235,7 +235,7 @@ where
     /// This function will attempt to write the entire contents of `buf` unless its capacity
     /// exceeds the one of the source buffer, in which case `GpuBuffer::capacity()` elements are written.
     pub fn write(&self, buf: &[T]) -> BufferResult<u64> {
-        let input_size = (buf.len() * std::mem::size_of::<T>()) as u64;
+        let input_size = std::mem::size_of_val(buf) as u64;
         let upload_size = if input_size > self.size {
             self.size
         } else {