diff --git a/Cargo.toml b/Cargo.toml index 9cba4f0..108be65 100644 --- a/Cargo.toml +++ b/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 } diff --git a/examples/image-compatibility/shader.wgsl b/examples/image-compatibility/shader.wgsl index 4b1040a..f97c6ea 100644 --- a/examples/image-compatibility/shader.wgsl +++ b/examples/image-compatibility/shader.wgsl @@ -7,7 +7,7 @@ fn main(@builtin(global_invocation_id) global_id: vec3) { let pixel = textureLoad(input, coord, 0); let dims = textureDimensions(input); - let mirror_coord = vec2(dims.x - coord.x, coord.y); + let mirror_coord = vec2(i32(dims.x) - coord.x, coord.y); textureStore(output, mirror_coord, pixel); } diff --git a/examples/mirror-image/shader.wgsl b/examples/mirror-image/shader.wgsl index 18a28b7..f938d47 100644 --- a/examples/mirror-image/shader.wgsl +++ b/examples/mirror-image/shader.wgsl @@ -7,7 +7,7 @@ fn main(@builtin(global_invocation_id) global_id: vec3) { let pixel = textureLoad(input, coord, 0); let dims = textureDimensions(input); - let mirror_coord = vec2(dims.x - coord.x, coord.y); + let mirror_coord = vec2(i32(dims.x) - coord.x, coord.y); textureStore(output, mirror_coord, pixel); } \ No newline at end of file diff --git a/examples/webcam/shader.wgsl b/examples/webcam/shader.wgsl index e09bf59..1da3d48 100644 --- a/examples/webcam/shader.wgsl +++ b/examples/webcam/shader.wgsl @@ -6,7 +6,7 @@ struct Time { @group(0) @binding(1) var output: texture_storage_2d; @group(0) @binding(2) var 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) { @@ -14,7 +14,7 @@ fn main(@builtin(global_invocation_id) global_id: vec3) { let pixel = textureLoad(input, coord, 0); let dims = textureDimensions(input); - let mirror_coord = vec2(dims.x - coord.x, coord.y); + let mirror_coord = vec2(i32(dims.x) - coord.x, coord.y); let t = pi * time.time; let colour = vec4(sin(t), sin(0.25 * t), sin(0.5 * t), 1.0); diff --git a/src/features/integrate_image.rs b/src/features/integrate_image.rs index af20ce4..167a923 100644 --- a/src/features/integrate_image.rs +++ b/src/features/integrate_image.rs @@ -112,12 +112,7 @@ where Container: std::ops::Deref, { 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

(mut bytes: Vec) -> Vec +fn bytes_to_primitive_vec

(mut bytes: Vec) -> Vec where P: image::Pixel, P::Subpixel: bytemuck::Pod, diff --git a/src/framework.rs b/src/framework.rs index fa6ce4b..53417c4 100644 --- a/src/framework.rs +++ b/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, ) diff --git a/src/kernel.rs b/src/kernel.rs index 5b8629b..a5870ff 100644 --- a/src/kernel.rs +++ b/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); diff --git a/src/primitives/buffers.rs b/src/primitives/buffers.rs index 8c321ee..c084cac 100644 --- a/src/primitives/buffers.rs +++ b/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::()) 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 { - let output_size = (buf.len() * std::mem::size_of::()) 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 { - let input_size = (buf.len() * std::mem::size_of::()) 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::()) 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 { - let input_size = (buf.len() * std::mem::size_of::()) as u64; + let input_size = std::mem::size_of_val(buf) as u64; let upload_size = if input_size > self.size { self.size } else { diff --git a/src/primitives/images.rs b/src/primitives/images.rs index 2c1df36..47ec84c 100644 --- a/src/primitives/images.rs +++ b/src/primitives/images.rs @@ -1,485 +1,479 @@ -use std::marker::PhantomData; - -use thiserror::Error; -use wgpu::util::DeviceExt; - -use crate::{GpuConstImage, GpuImage}; - -use super::{ImgOps, PixelInfo}; - -// TODO https://github.com/bitflags/bitflags/issues/180 -const GPU_IMAGE_USAGES: wgpu::TextureUsages = wgpu::TextureUsages::from_bits_truncate( - wgpu::TextureUsages::STORAGE_BINDING.bits() - | wgpu::TextureUsages::COPY_SRC.bits() - | wgpu::TextureUsages::COPY_DST.bits(), -); -const GPU_CONST_IMAGE_USAGES: wgpu::TextureUsages = wgpu::TextureUsages::from_bits_truncate( - wgpu::TextureUsages::TEXTURE_BINDING.bits() | wgpu::TextureUsages::COPY_DST.bits(), -); - -#[derive(Error, Debug)] -pub enum ImageOutputError { - #[error(transparent)] - BufferError(#[from] crate::primitives::buffers::BufferError), - #[error( - "Output is too small (required size {required} bytes, current size {current} bytes). " - )] - BufferTooSmall { required: usize, current: usize }, -} - -#[derive(Error, Debug)] -pub enum ImageInputError { - #[error("Input does not contains an integer number of pixels.")] - NotIntegerPixelNumber, - #[error("Input does not contains an integer number of rows.")] - NotIntegerRowNumber, -} - -impl<'fw, P> ImgOps<'fw> for GpuImage<'fw, P> -where - P: PixelInfo, -{ - fn as_binding_resource(&self) -> wgpu::BindingResource { - wgpu::BindingResource::TextureView(&self.full_view) - } - - fn as_gpu_texture(&self) -> &wgpu::Texture { - &self.texture - } - - fn get_wgpu_extent3d(&self) -> wgpu::Extent3d { - self.size - } - - fn dimensions(&self) -> (u32, u32) { - (self.size.width, self.size.height) - } - - fn new(fw: &'fw crate::Framework, width: u32, height: u32) -> Self { - let size = wgpu::Extent3d { - width, - height, - depth_or_array_layers: 1, - }; - - let format = P::wgpu_format(); - - let texture = fw.device.create_texture(&wgpu::TextureDescriptor { - label: Some("GpuImage::new"), - size, - dimension: wgpu::TextureDimension::D2, - mip_level_count: 1, - sample_count: 1, - format, - usage: GPU_IMAGE_USAGES, - }); - - let full_view = texture.create_view(&wgpu::TextureViewDescriptor::default()); - - Self { - fw, - texture, - size, - full_view, - pixel: PhantomData, - } - } - - fn from_bytes(fw: &'fw crate::Framework, data: &[u8], width: u32, height: u32) -> Self { - let size = wgpu::Extent3d { - width, - height, - depth_or_array_layers: 1, - }; - - let format = P::wgpu_format(); - - let texture = fw.device.create_texture_with_data( - &fw.queue, - &wgpu::TextureDescriptor { - label: Some("GpuImage::from_bytes"), - size, - dimension: wgpu::TextureDimension::D2, - mip_level_count: 1, - sample_count: 1, - format, - usage: GPU_IMAGE_USAGES, - }, - data, - ); - - let full_view = texture.create_view(&Default::default()); - - Self { - fw, - texture, - size, - full_view, - pixel: PhantomData, - } - } - - /// Constructs an image from a [`wgpu::Texture`] and its [`wgpu::Extent3d`]. - /// - /// # Safety - /// If any of the following conditions are not satisfied, the image will - /// panic at any time during its usage. - /// - `texture` needs to be `wgpu::TextureUsages::STORAGE_BINDING`, `wgpu::TextureUsages::COPY_SRC`, - /// and `wgpu::TextureUsages::COPY_DST`` usable. - /// - `T` needs to be the exact same codification `texture` is. - /// - `dimensions` needs to have the exact `width` and `height` of `texture` and `depth_or_array_layers = 1` - fn from_gpu_parts( - fw: &'fw crate::Framework, - texture: wgpu::Texture, - dimensions: wgpu::Extent3d, - ) -> Self { - let full_view = texture.create_view(&Default::default()); - - Self { - fw, - texture, - size: dimensions, - full_view, - pixel: PhantomData, - } - } - - fn into_gpu_parts(self) -> (wgpu::Texture, wgpu::Extent3d) { - (self.texture, self.size) - } -} - -impl<'fw, P> GpuImage<'fw, P> -where - P: PixelInfo, -{ - /// Pulls some elements from the [`GpuImage`] into `buf`, returning how many pixels were read. - pub async fn read(&self, buf: &mut [u8]) -> Result { - use std::num::NonZeroU32; - - let (width, height) = self.dimensions(); - - let img_bytes = (width * height) as usize * P::byte_size(); - - if buf.len() < img_bytes { - return Err(ImageOutputError::BufferTooSmall { - required: img_bytes, - current: buf.len(), - }); - } - - let bytes_per_pixel = P::byte_size() as u32; - let unpadded_bytes_per_row = self.size.width * bytes_per_pixel; - let align = wgpu::COPY_BYTES_PER_ROW_ALIGNMENT; - let padded_bytes_per_row_padding = (align - unpadded_bytes_per_row % align) % align; - let padded_bytes_per_row = unpadded_bytes_per_row + padded_bytes_per_row_padding; - - let staging_size = (padded_bytes_per_row * self.size.height) as usize; - - let staging = self.fw.device.create_buffer(&wgpu::BufferDescriptor { - label: Some("GpuImage::read"), - size: staging_size as u64, - usage: wgpu::BufferUsages::COPY_SRC | wgpu::BufferUsages::COPY_DST, - mapped_at_creation: false, - }); - - let mut encoder = self - .fw - .device - .create_command_encoder(&wgpu::CommandEncoderDescriptor { - label: Some("GpuImage::read"), - }); - - let copy_texture = wgpu::ImageCopyTexture { - aspect: wgpu::TextureAspect::All, - mip_level: 0, - origin: wgpu::Origin3d::ZERO, - texture: &self.texture, - }; - - let copy_buffer = wgpu::ImageCopyBuffer { - buffer: &staging, - layout: wgpu::ImageDataLayout { - offset: 0, - bytes_per_row: NonZeroU32::new(padded_bytes_per_row), - rows_per_image: None, - }, - }; - - encoder.copy_texture_to_buffer(copy_texture, copy_buffer, self.size); - - self.fw.queue.submit(Some(encoder.finish())); - - let (tx, rx) = futures::channel::oneshot::channel(); - wgpu::util::DownloadBuffer::read_buffer( - &self.fw.device, - &self.fw.queue, - &staging.slice(..), - move |result| { - tx.send(result) - .unwrap_or_else(|_| panic!("Failed to download buffer.")); - }, - ); - let download = rx.await.unwrap().unwrap(); - - let bytes_read: usize = download - .chunks(padded_bytes_per_row as usize) - .zip(buf.chunks_mut(unpadded_bytes_per_row as usize)) - .map(|(src, dest)| { - dest.copy_from_slice(&src[0..unpadded_bytes_per_row as usize]); - dest.len() - }) - .sum(); - - Ok(bytes_read / P::byte_size()) - } - - /// Pulls all the pixels from the [`GpuImage`] into a [`Vec`]. - pub async fn read_vec(&self) -> Result, ImageOutputError> { - let (width, height) = self.dimensions(); - let img_pixels = (width * height) as usize * P::byte_size(); - - let mut buf = vec![0u8; img_pixels]; - self.read(&mut buf).await?; - - Ok(buf) - } - - /// Blocking version of `GpuImage::read()`. - pub fn read_blocking(&self, buf: &mut [u8]) -> Result { - futures::executor::block_on(self.read(buf)) - } - - /// Blocking version of `GpuImage::read_vec()`. - pub fn read_vec_blocking(&self) -> Result, ImageOutputError> { - futures::executor::block_on(self.read_vec()) - } - - /// Writes a buffer into this [`GpuImage`], returning how many pixels were written. The operation is instantly offloaded. - /// - /// This function will attempt to write the entire contents of `buf`, unless its capacity - /// exceeds the one of the image, in which case the first `width * height` pixels are written. - pub fn write(&self, buf: &[u8]) -> Result { - use std::num::NonZeroU32; - - if buf.len() % P::byte_size() != 0 { - return Err(ImageInputError::NotIntegerPixelNumber); - } - if buf.len() % (P::byte_size() * self.size.width as usize) != 0 { - return Err(ImageInputError::NotIntegerRowNumber); - } - - let image_bytes = P::byte_size() * (self.size.width * self.size.height) as usize; - - let (write_buf, size) = match buf.len().cmp(&image_bytes) { - std::cmp::Ordering::Less => ( - buf, - wgpu::Extent3d { - width: self.size.width, - height: (buf.len() / P::byte_size()) as u32 / self.size.width, - depth_or_array_layers: 1, - }, - ), - std::cmp::Ordering::Equal => (buf, self.size), - std::cmp::Ordering::Greater => (&buf[..image_bytes], self.size), - }; - - self.fw.queue.write_texture( - self.texture.as_image_copy(), - write_buf, - wgpu::ImageDataLayout { - offset: 0, - bytes_per_row: Some( - NonZeroU32::new(P::byte_size() as u32 * self.size.width) - .expect("Could not create a NonZeroU32."), - ), - rows_per_image: None, - }, - size, - ); - - let encoder = self - .fw - .device - .create_command_encoder(&wgpu::CommandEncoderDescriptor { - label: Some("GpuImage::write"), - }); - - self.fw.queue.submit(Some(encoder.finish())); - - Ok((size.width * size.height) as usize) - } -} - -impl<'fw, P> ImgOps<'fw> for GpuConstImage<'fw, P> -where - P: PixelInfo, -{ - fn as_binding_resource(&self) -> wgpu::BindingResource { - wgpu::BindingResource::TextureView(&self.full_view) - } - - fn as_gpu_texture(&self) -> &wgpu::Texture { - &self.texture - } - - fn get_wgpu_extent3d(&self) -> wgpu::Extent3d { - self.size - } - - fn dimensions(&self) -> (u32, u32) { - (self.size.width, self.size.height) - } - - fn new(fw: &'fw crate::Framework, width: u32, height: u32) -> Self { - let size = wgpu::Extent3d { - width, - height, - depth_or_array_layers: 1, - }; - - let format = P::wgpu_format(); - - let texture = fw.device.create_texture(&wgpu::TextureDescriptor { - label: Some("GpuConstImage::new"), - size, - dimension: wgpu::TextureDimension::D2, - mip_level_count: 1, - sample_count: 1, - format, - usage: GPU_CONST_IMAGE_USAGES, - }); - - let full_view = texture.create_view(&wgpu::TextureViewDescriptor::default()); - - Self { - fw, - texture, - size, - full_view, - pixel: PhantomData, - } - } - - fn from_bytes(fw: &'fw crate::Framework, data: &[u8], width: u32, height: u32) -> Self { - let size = wgpu::Extent3d { - width, - height, - depth_or_array_layers: 1, - }; - - let format = P::wgpu_format(); - - let texture = fw.device.create_texture_with_data( - &fw.queue, - &wgpu::TextureDescriptor { - label: Some("GpuConstImage::from_bytes"), - size, - dimension: wgpu::TextureDimension::D2, - mip_level_count: 1, - sample_count: 1, - format, - usage: GPU_CONST_IMAGE_USAGES, - }, - data, - ); - - let full_view = texture.create_view(&Default::default()); - - Self { - fw, - texture, - size, - full_view, - pixel: PhantomData, - } - } - - /// Constructs an image from a [`wgpu::Texture`] and its [`wgpu::Extent3d`]. - /// - /// # Safety - /// If any of the following conditions are not satisfied, the image will - /// panic at any time during its usage. - /// - `texture` needs to be `wgpu::TextureUsages::TEXTURE_BINDING` and `wgpu::TextureUsages::COPY_SRC` usable. - /// - `T` needs to be the exact same codification `texture` is. - /// - `dimensions` needs to have the exact `width` and `height` of `texture` and `depth_or_array_layers = 1` - fn from_gpu_parts( - fw: &'fw crate::Framework, - texture: wgpu::Texture, - dimensions: wgpu::Extent3d, - ) -> Self { - let full_view = texture.create_view(&Default::default()); - - Self { - fw, - texture, - size: dimensions, - full_view, - pixel: PhantomData, - } - } - - fn into_gpu_parts(self) -> (wgpu::Texture, wgpu::Extent3d) { - (self.texture, self.size) - } -} - -impl<'fw, P> GpuConstImage<'fw, P> -where - P: PixelInfo, -{ - /// Writes a buffer into this [`GpuConstImage`], returning how many pixels were written. The operation is instantly offloaded. - /// - /// This function will attempt to write the entire contents of `buf`, unless its capacity - /// exceeds the one of the image, in which case the first `width * height` pixels are written. - pub fn write(&self, buf: &[u8]) -> Result { - use std::num::NonZeroU32; - - if buf.len() % P::byte_size() != 0 { - return Err(ImageInputError::NotIntegerPixelNumber); - } - if buf.len() % (P::byte_size() * self.size.width as usize) != 0 { - return Err(ImageInputError::NotIntegerRowNumber); - } - - let image_bytes = P::byte_size() * (self.size.width * self.size.height) as usize; - - let (write_buf, size) = match buf.len().cmp(&image_bytes) { - std::cmp::Ordering::Less => ( - buf, - wgpu::Extent3d { - width: self.size.width, - height: buf.len() as u32 / self.size.width, - depth_or_array_layers: 1, - }, - ), - std::cmp::Ordering::Equal => (buf, self.size), - std::cmp::Ordering::Greater => (&buf[..image_bytes], self.size), - }; - - self.fw.queue.write_texture( - self.texture.as_image_copy(), - write_buf, - wgpu::ImageDataLayout { - offset: 0, - bytes_per_row: Some( - NonZeroU32::new(P::byte_size() as u32 * self.size.width) - .expect("Could not create a NonZeroU32."), - ), - rows_per_image: None, - }, - size, - ); - - let encoder = self - .fw - .device - .create_command_encoder(&wgpu::CommandEncoderDescriptor { - label: Some("GpuConstImage::write"), - }); - - self.fw.queue.submit(Some(encoder.finish())); - - Ok((size.width * size.height) as usize) - } -} +use std::marker::PhantomData; + +use thiserror::Error; +use wgpu::util::DeviceExt; + +use crate::{GpuConstImage, GpuImage}; + +use super::{ImgOps, PixelInfo}; + +// TODO https://github.com/bitflags/bitflags/issues/180 +const GPU_IMAGE_USAGES: wgpu::TextureUsages = wgpu::TextureUsages::from_bits_truncate( + wgpu::TextureUsages::STORAGE_BINDING.bits() + | wgpu::TextureUsages::COPY_SRC.bits() + | wgpu::TextureUsages::COPY_DST.bits(), +); +const GPU_CONST_IMAGE_USAGES: wgpu::TextureUsages = wgpu::TextureUsages::from_bits_truncate( + wgpu::TextureUsages::TEXTURE_BINDING.bits() | wgpu::TextureUsages::COPY_DST.bits(), +); + +#[derive(Error, Debug)] +pub enum ImageOutputError { + #[error(transparent)] + BufferError(#[from] crate::primitives::buffers::BufferError), + #[error( + "Output is too small (required size {required} bytes, current size {current} bytes). " + )] + BufferTooSmall { required: usize, current: usize }, +} + +#[derive(Error, Debug)] +pub enum ImageInputError { + #[error("Input does not contains an integer number of pixels.")] + NotIntegerPixelNumber, + #[error("Input does not contains an integer number of rows.")] + NotIntegerRowNumber, +} + +impl<'fw, P> ImgOps<'fw> for GpuImage<'fw, P> +where + P: PixelInfo, +{ + fn as_binding_resource(&self) -> wgpu::BindingResource { + wgpu::BindingResource::TextureView(&self.full_view) + } + + fn as_gpu_texture(&self) -> &wgpu::Texture { + &self.texture + } + + fn get_wgpu_extent3d(&self) -> wgpu::Extent3d { + self.size + } + + fn dimensions(&self) -> (u32, u32) { + (self.size.width, self.size.height) + } + + fn new(fw: &'fw crate::Framework, width: u32, height: u32) -> Self { + let size = wgpu::Extent3d { + width, + height, + depth_or_array_layers: 1, + }; + + let format = P::wgpu_format(); + + let texture = fw.device.create_texture(&wgpu::TextureDescriptor { + label: Some("GpuImage::new"), + size, + dimension: wgpu::TextureDimension::D2, + mip_level_count: 1, + sample_count: 1, + format, + usage: GPU_IMAGE_USAGES, + view_formats: &[format], + }); + + let full_view = texture.create_view(&wgpu::TextureViewDescriptor::default()); + + Self { + fw, + texture, + size, + full_view, + pixel: PhantomData, + } + } + + fn from_bytes(fw: &'fw crate::Framework, data: &[u8], width: u32, height: u32) -> Self { + let size = wgpu::Extent3d { + width, + height, + depth_or_array_layers: 1, + }; + + let format = P::wgpu_format(); + + let texture = fw.device.create_texture_with_data( + &fw.queue, + &wgpu::TextureDescriptor { + label: Some("GpuImage::from_bytes"), + size, + dimension: wgpu::TextureDimension::D2, + mip_level_count: 1, + sample_count: 1, + format, + usage: GPU_IMAGE_USAGES, + view_formats: &[format], + }, + wgpu::util::TextureDataOrder::LayerMajor, + data, + ); + + let full_view = texture.create_view(&Default::default()); + + Self { + fw, + texture, + size, + full_view, + pixel: PhantomData, + } + } + + /// Constructs an image from a [`wgpu::Texture`] and its [`wgpu::Extent3d`]. + /// + /// # Safety + /// If any of the following conditions are not satisfied, the image will + /// panic at any time during its usage. + /// - `texture` needs to be `wgpu::TextureUsages::STORAGE_BINDING`, `wgpu::TextureUsages::COPY_SRC`, + /// and `wgpu::TextureUsages::COPY_DST`` usable. + /// - `T` needs to be the exact same codification `texture` is. + /// - `dimensions` needs to have the exact `width` and `height` of `texture` and `depth_or_array_layers = 1` + fn from_gpu_parts( + fw: &'fw crate::Framework, + texture: wgpu::Texture, + dimensions: wgpu::Extent3d, + ) -> Self { + let full_view = texture.create_view(&Default::default()); + + Self { + fw, + texture, + size: dimensions, + full_view, + pixel: PhantomData, + } + } + + fn into_gpu_parts(self) -> (wgpu::Texture, wgpu::Extent3d) { + (self.texture, self.size) + } +} + +impl<'fw, P> GpuImage<'fw, P> +where + P: PixelInfo, +{ + /// Pulls some elements from the [`GpuImage`] into `buf`, returning how many pixels were read. + pub async fn read(&self, buf: &mut [u8]) -> Result { + let (width, height) = self.dimensions(); + + let img_bytes = (width * height) as usize * P::byte_size(); + + if buf.len() < img_bytes { + return Err(ImageOutputError::BufferTooSmall { + required: img_bytes, + current: buf.len(), + }); + } + + let bytes_per_pixel = P::byte_size() as u32; + let unpadded_bytes_per_row = self.size.width * bytes_per_pixel; + let align = wgpu::COPY_BYTES_PER_ROW_ALIGNMENT; + let padded_bytes_per_row_padding = (align - unpadded_bytes_per_row % align) % align; + let padded_bytes_per_row = unpadded_bytes_per_row + padded_bytes_per_row_padding; + + let staging_size = (padded_bytes_per_row * self.size.height) as usize; + + let staging = self.fw.device.create_buffer(&wgpu::BufferDescriptor { + label: Some("GpuImage::read"), + size: staging_size as u64, + usage: wgpu::BufferUsages::COPY_SRC | wgpu::BufferUsages::COPY_DST, + mapped_at_creation: false, + }); + + let mut encoder = self + .fw + .device + .create_command_encoder(&wgpu::CommandEncoderDescriptor { + label: Some("GpuImage::read"), + }); + + let copy_texture = wgpu::ImageCopyTexture { + aspect: wgpu::TextureAspect::All, + mip_level: 0, + origin: wgpu::Origin3d::ZERO, + texture: &self.texture, + }; + + let copy_buffer = wgpu::ImageCopyBuffer { + buffer: &staging, + layout: wgpu::ImageDataLayout { + offset: 0, + bytes_per_row: Option::Some(padded_bytes_per_row), + rows_per_image: None, + }, + }; + + encoder.copy_texture_to_buffer(copy_texture, copy_buffer, self.size); + + self.fw.queue.submit(Some(encoder.finish())); + + let (tx, rx) = futures::channel::oneshot::channel(); + wgpu::util::DownloadBuffer::read_buffer( + &self.fw.device, + &self.fw.queue, + &staging.slice(..), + move |result| { + tx.send(result) + .unwrap_or_else(|_| panic!("Failed to download buffer.")); + }, + ); + let download = rx.await.unwrap().unwrap(); + + let bytes_read: usize = download + .chunks(padded_bytes_per_row as usize) + .zip(buf.chunks_mut(unpadded_bytes_per_row as usize)) + .map(|(src, dest)| { + dest.copy_from_slice(&src[0..unpadded_bytes_per_row as usize]); + dest.len() + }) + .sum(); + + Ok(bytes_read / P::byte_size()) + } + + /// Pulls all the pixels from the [`GpuImage`] into a [`Vec`]. + pub async fn read_vec(&self) -> Result, ImageOutputError> { + let (width, height) = self.dimensions(); + let img_pixels = (width * height) as usize * P::byte_size(); + + let mut buf = vec![0u8; img_pixels]; + self.read(&mut buf).await?; + + Ok(buf) + } + + /// Blocking version of `GpuImage::read()`. + pub fn read_blocking(&self, buf: &mut [u8]) -> Result { + futures::executor::block_on(self.read(buf)) + } + + /// Blocking version of `GpuImage::read_vec()`. + pub fn read_vec_blocking(&self) -> Result, ImageOutputError> { + futures::executor::block_on(self.read_vec()) + } + + /// Writes a buffer into this [`GpuImage`], returning how many pixels were written. The operation is instantly offloaded. + /// + /// This function will attempt to write the entire contents of `buf`, unless its capacity + /// exceeds the one of the image, in which case the first `width * height` pixels are written. + pub fn write(&self, buf: &[u8]) -> Result { + if buf.len() % P::byte_size() != 0 { + return Err(ImageInputError::NotIntegerPixelNumber); + } + if buf.len() % (P::byte_size() * self.size.width as usize) != 0 { + return Err(ImageInputError::NotIntegerRowNumber); + } + + let image_bytes = P::byte_size() * (self.size.width * self.size.height) as usize; + + let (write_buf, size) = match buf.len().cmp(&image_bytes) { + std::cmp::Ordering::Less => ( + buf, + wgpu::Extent3d { + width: self.size.width, + height: (buf.len() / P::byte_size()) as u32 / self.size.width, + depth_or_array_layers: 1, + }, + ), + std::cmp::Ordering::Equal => (buf, self.size), + std::cmp::Ordering::Greater => (&buf[..image_bytes], self.size), + }; + + self.fw.queue.write_texture( + self.texture.as_image_copy(), + write_buf, + wgpu::ImageDataLayout { + offset: 0, + bytes_per_row: Some(P::byte_size() as u32 * self.size.width), + rows_per_image: None, + }, + size, + ); + + let encoder = self + .fw + .device + .create_command_encoder(&wgpu::CommandEncoderDescriptor { + label: Some("GpuImage::write"), + }); + + self.fw.queue.submit(Some(encoder.finish())); + + Ok((size.width * size.height) as usize) + } +} + +impl<'fw, P> ImgOps<'fw> for GpuConstImage<'fw, P> +where + P: PixelInfo, +{ + fn as_binding_resource(&self) -> wgpu::BindingResource { + wgpu::BindingResource::TextureView(&self.full_view) + } + + fn as_gpu_texture(&self) -> &wgpu::Texture { + &self.texture + } + + fn get_wgpu_extent3d(&self) -> wgpu::Extent3d { + self.size + } + + fn dimensions(&self) -> (u32, u32) { + (self.size.width, self.size.height) + } + + fn new(fw: &'fw crate::Framework, width: u32, height: u32) -> Self { + let size = wgpu::Extent3d { + width, + height, + depth_or_array_layers: 1, + }; + + let format = P::wgpu_format(); + + let texture = fw.device.create_texture(&wgpu::TextureDescriptor { + label: Some("GpuConstImage::new"), + size, + dimension: wgpu::TextureDimension::D2, + mip_level_count: 1, + sample_count: 1, + format, + usage: GPU_CONST_IMAGE_USAGES, + view_formats: &[format], + }); + + let full_view = texture.create_view(&wgpu::TextureViewDescriptor::default()); + + Self { + fw, + texture, + size, + full_view, + pixel: PhantomData, + } + } + + fn from_bytes(fw: &'fw crate::Framework, data: &[u8], width: u32, height: u32) -> Self { + let size = wgpu::Extent3d { + width, + height, + depth_or_array_layers: 1, + }; + + let format = P::wgpu_format(); + + let texture = fw.device.create_texture_with_data( + &fw.queue, + &wgpu::TextureDescriptor { + label: Some("GpuConstImage::from_bytes"), + size, + dimension: wgpu::TextureDimension::D2, + mip_level_count: 1, + sample_count: 1, + format, + usage: GPU_CONST_IMAGE_USAGES, + view_formats: &[format], + }, + wgpu::util::TextureDataOrder::LayerMajor, + data, + ); + + let full_view = texture.create_view(&Default::default()); + + Self { + fw, + texture, + size, + full_view, + pixel: PhantomData, + } + } + + /// Constructs an image from a [`wgpu::Texture`] and its [`wgpu::Extent3d`]. + /// + /// # Safety + /// If any of the following conditions are not satisfied, the image will + /// panic at any time during its usage. + /// - `texture` needs to be `wgpu::TextureUsages::TEXTURE_BINDING` and `wgpu::TextureUsages::COPY_SRC` usable. + /// - `T` needs to be the exact same codification `texture` is. + /// - `dimensions` needs to have the exact `width` and `height` of `texture` and `depth_or_array_layers = 1` + fn from_gpu_parts( + fw: &'fw crate::Framework, + texture: wgpu::Texture, + dimensions: wgpu::Extent3d, + ) -> Self { + let full_view = texture.create_view(&Default::default()); + + Self { + fw, + texture, + size: dimensions, + full_view, + pixel: PhantomData, + } + } + + fn into_gpu_parts(self) -> (wgpu::Texture, wgpu::Extent3d) { + (self.texture, self.size) + } +} + +impl<'fw, P> GpuConstImage<'fw, P> +where + P: PixelInfo, +{ + /// Writes a buffer into this [`GpuConstImage`], returning how many pixels were written. The operation is instantly offloaded. + /// + /// This function will attempt to write the entire contents of `buf`, unless its capacity + /// exceeds the one of the image, in which case the first `width * height` pixels are written. + pub fn write(&self, buf: &[u8]) -> Result { + if buf.len() % P::byte_size() != 0 { + return Err(ImageInputError::NotIntegerPixelNumber); + } + if buf.len() % (P::byte_size() * self.size.width as usize) != 0 { + return Err(ImageInputError::NotIntegerRowNumber); + } + + let image_bytes = P::byte_size() * (self.size.width * self.size.height) as usize; + + let (write_buf, size) = match buf.len().cmp(&image_bytes) { + std::cmp::Ordering::Less => ( + buf, + wgpu::Extent3d { + width: self.size.width, + height: buf.len() as u32 / self.size.width, + depth_or_array_layers: 1, + }, + ), + std::cmp::Ordering::Equal => (buf, self.size), + std::cmp::Ordering::Greater => (&buf[..image_bytes], self.size), + }; + + self.fw.queue.write_texture( + self.texture.as_image_copy(), + write_buf, + wgpu::ImageDataLayout { + offset: 0, + bytes_per_row: Some(P::byte_size() as u32 * self.size.width), + rows_per_image: None, + }, + size, + ); + + let encoder = self + .fw + .device + .create_command_encoder(&wgpu::CommandEncoderDescriptor { + label: Some("GpuConstImage::write"), + }); + + self.fw.queue.submit(Some(encoder.finish())); + + Ok((size.width * size.height) as usize) + } +} diff --git a/src/primitives/samplers.rs b/src/primitives/samplers.rs index 310ba29..037fbea 100644 --- a/src/primitives/samplers.rs +++ b/src/primitives/samplers.rs @@ -28,7 +28,7 @@ impl Sampler { lod_min_clamp: 0.0, lod_max_clamp: std::f32::MAX, compare: None, - anisotropy_clamp: None, + anisotropy_clamp: 1, border_color: None, }); Self {