Merge branch 'trunk' into naga-ray-tracing-pipelines

Author: Vecvec

.github/workflows/changelog.yml

@@ -6,6 +6,12 @@ on:
       - ".github/workflows/changelog.yml"
       - "CHANGELOG.md"
       - "xtask/**/*"
+    types:
+      - opened
+      - synchronize
+      - reopened
+      - labeled
+      - unlabeled
 
 env:
   #
@@ -38,6 +44,7 @@ jobs:
         with:
           fetch-depth: 0
 
+      # NOTE: Keep label name(s) in sync. with `xtask`'s implementation.
       - name: Run `cargo xtask changelog …`
         run: |
-          cargo xtask changelog "origin/${{ github.event.pull_request.base.ref }}"
+          cargo xtask changelog "origin/${{ github.event.pull_request.base.ref }}" ${{ contains(github.event.pull_request.labels.*.name, 'changelog: released entry changed') && '--allow-released-changes' || ''  }}

CHANGELOG.md

@@ -110,6 +110,7 @@ By @SupaMaggie70Incorporated in [#8206](https://github.com/gfx-rs/wgpu/pull/8206
 
 - Added support for transient textures on Vulkan and Metal. By @opstic in [#8247](https://github.com/gfx-rs/wgpu/pull/8247)
 - Implement shader triangle barycentric coordinate builtins. By @atlv24 in [#8320](https://github.com/gfx-rs/wgpu/pull/8320).
+- Added support for binding arrays of storage textures on Metal. By @msvbg in [#8464](https://github.com/gfx-rs/wgpu/pull/8464)
 
 ### Changes
 
@@ -171,6 +172,7 @@ By @SupaMaggie70Incorporated in [#8206](https://github.com/gfx-rs/wgpu/pull/8206
 #### GLES
 
 - Fix race when downloading texture from compute shader pass. By @SpeedCrash100 in [#8527](https://github.com/gfx-rs/wgpu/pull/8527)
+- Fix double window class registration when dynamic libraries are used. By @Azorlogh in [#8548](https://github.com/gfx-rs/wgpu/pull/8548)
 
 #### hal
 

naga/src/back/msl/writer.rs

@@ -7152,9 +7152,11 @@ template <typename A>
                                         }
                                     }
                                     crate::ImageClass::Storage { .. } => {
-                                        return Err(Error::UnsupportedArrayOf(
-                                            "read-write textures".to_string(),
-                                        ));
+                                        if options.lang_version < (3, 0) {
+                                            return Err(Error::UnsupportedArrayOf(
+                                                "read-write textures".to_string(),
+                                            ));
+                                        }
                                     }
                                     crate::ImageClass::External => {
                                         return Err(Error::UnsupportedArrayOf(

naga/tests/naga/validation.rs

@@ -1100,3 +1100,55 @@ fn used() {
         Some("_ = arr2[3];"),
     );
 }
+
+/// Expects parsing `input` to succeed and its validation to fail with error equal to `snapshot`.
+#[track_caller]
+fn check_wgsl_validation_error_message(input: &str, snapshot: &str) {
+    let module = naga::front::wgsl::parse_str(input).unwrap();
+    let err = valid::Validator::new(Default::default(), valid::Capabilities::all())
+        .validate(&module)
+        .expect_err("module should be invalid")
+        .emit_to_string(input);
+    if err != snapshot {
+        for diff in diff::lines(snapshot, &err) {
+            match diff {
+                diff::Result::Left(l) => println!("-{l}"),
+                diff::Result::Both(l, _) => println!(" {l}"),
+                diff::Result::Right(r) => println!("+{r}"),
+            }
+        }
+        panic!("Error does not match the expected snapshot");
+    }
+}
+
+#[test]
+fn image_store_type_mismatch() {
+    check_wgsl_validation_error_message(
+        r#"
+@group(0) @binding(0)
+var input_texture: texture_depth_2d;
+@group(0) @binding(1)
+var input_sampler: sampler;
+@group(0) @binding(2)
+var output_texture: texture_storage_2d<r32float,write>;
+
+@compute @workgroup_size(1, 1)
+fn main() {
+    let d: vec4<f32> = textureGather(input_texture, input_sampler, vec2f(0.0));
+    let min_d = min(min(d[0], d[1]), min(d[2], d[3]));
+    textureStore(output_texture, vec2u(1), min_d);
+}
+"#,
+        r#"error: Entry point main at Compute is invalid
+   ┌─ wgsl:12:17
+   │
+12 │     let min_d = min(min(d[0], d[1]), min(d[2], d[3]));
+   │                 ^^^ this value is of type Scalar(Scalar { kind: Float, width: 4 })
+13 │     textureStore(output_texture, vec2u(1), min_d);
+   │     ^^^^^^^^^^^^ expects a value argument of type Vector { size: Quad, scalar: Scalar { kind: Float, width: 4 } }
+   │
+   = Image store value parameter type mismatch
+
+"#,
+    );
+}

tests/tests/wgpu-gpu/binding_array/storage_textures.rs

@@ -2,8 +2,8 @@ use std::num::NonZeroU32;
 
 use wgpu::*;
 use wgpu_test::{
-    gpu_test, image::ReadbackBuffers, FailureCase, GpuTestConfiguration, GpuTestInitializer,
-    TestParameters, TestingContext,
+    gpu_test, image::ReadbackBuffers, GpuTestConfiguration, GpuTestInitializer, TestParameters,
+    TestingContext,
 };
 
 pub fn all_tests(tests: &mut Vec<GpuTestInitializer>) {
@@ -26,8 +26,7 @@ static BINDING_ARRAY_STORAGE_TEXTURES: GpuTestConfiguration = GpuTestConfigurati
             .limits(Limits {
                 max_binding_array_elements_per_shader_stage: 17,
                 ..Limits::default()
-            })
-            .expect_fail(FailureCase::backend(Backends::METAL)),
+            }),
     )
     .run_async(|ctx| async move { binding_array_storage_textures(ctx, false).await });
 
@@ -45,8 +44,7 @@ static PARTIAL_BINDING_ARRAY_STORAGE_TEXTURES: GpuTestConfiguration = GpuTestCon
             .limits(Limits {
                 max_binding_array_elements_per_shader_stage: 33,
                 ..Limits::default()
-            })
-            .expect_fail(FailureCase::backend(Backends::METAL)),
+            }),
     )
     .run_async(|ctx| async move { binding_array_storage_textures(ctx, true).await });
 

wgpu-core/src/command/clear.rs

@@ -189,6 +189,8 @@ pub(super) fn clear_buffer(
         });
     }
 
+    // This must happen after parameter validation (so that errors are reported
+    // as required by the spec), but before any side effects.
     if offset == end_offset {
         log::trace!("Ignoring fill_buffer of size 0");
         return Ok(());

wgpu-core/src/command/transfer.rs

@@ -1037,6 +1037,8 @@ pub(super) fn copy_buffer_to_buffer(
         .into());
     }
 
+    // This must happen after parameter validation (so that errors are reported
+    // as required by the spec), but before any side effects.
     if size == 0 {
         log::trace!("Ignoring copy_buffer_to_buffer of size 0");
         return Ok(());
@@ -1255,6 +1257,8 @@ pub(super) fn copy_texture_to_buffer(
         .check_usage(BufferUsages::COPY_DST)
         .map_err(TransferError::MissingBufferUsage)?;
 
+    // This must happen after parameter validation (so that errors are reported
+    // as required by the spec), but before any side effects.
     if copy_size.width == 0 || copy_size.height == 0 || copy_size.depth_or_array_layers == 0 {
         log::trace!("Ignoring copy_texture_to_buffer of size 0");
         return Ok(());
@@ -1376,12 +1380,6 @@ pub(super) fn copy_texture_to_texture(
         .into());
     }
 
-    // Handle texture init *before* dealing with barrier transitions so we
-    // have an easier time inserting "immediate-inits" that may be required
-    // by prior discards in rare cases.
-    handle_src_texture_init(state, source, copy_size, src_texture)?;
-    handle_dst_texture_init(state, destination, copy_size, dst_texture)?;
-
     let src_raw = src_texture.try_raw(state.snatch_guard)?;
     src_texture
         .check_usage(TextureUsages::COPY_SRC)
@@ -1391,11 +1389,19 @@ pub(super) fn copy_texture_to_texture(
         .check_usage(TextureUsages::COPY_DST)
         .map_err(TransferError::MissingTextureUsage)?;
 
+    // This must happen after parameter validation (so that errors are reported
+    // as required by the spec), but before any side effects.
     if copy_size.width == 0 || copy_size.height == 0 || copy_size.depth_or_array_layers == 0 {
         log::trace!("Ignoring copy_texture_to_texture of size 0");
         return Ok(());
     }
 
+    // Handle texture init *before* dealing with barrier transitions so we
+    // have an easier time inserting "immediate-inits" that may be required
+    // by prior discards in rare cases.
+    handle_src_texture_init(state, source, copy_size, src_texture)?;
+    handle_dst_texture_init(state, destination, copy_size, dst_texture)?;
+
     let src_pending =
         state
             .tracker

wgpu-core/src/conv.rs

@@ -199,6 +199,17 @@ pub fn map_texture_usage_from_hal(uses: wgt::TextureUses) -> wgt::TextureUsages
     u
 }
 
+/// Check the requested texture size against the supported limits.
+///
+/// This function implements the texture size and sample count checks in [vtd
+/// dimension step]. The format checks are elsewhere in [`create_texture`]`.
+///
+/// Note that while there is some basic checking of the sample count here, there
+/// is an additional set of checks when `sample_count > 1` elsewhere in
+/// [`create_texture`]`.
+///
+/// [vtd dimension step]: https://www.w3.org/TR/2025/CRD-webgpu-20251120/#:~:text=or%204.-,If%20descriptor.dimension%20is
+/// [`create_texture`]: crate::device::Device::create_texture
 pub fn check_texture_dimension_size(
     dimension: wgt::TextureDimension,
     wgt::Extent3d {

wgpu-core/src/device/queue.rs

@@ -537,6 +537,8 @@ impl Queue {
         let data_size = if let Some(data_size) = wgt::BufferSize::new(data_size) {
             data_size
         } else {
+            // This must happen after parameter validation (so that errors are reported
+            // as required by the spec), but before any side effects.
             log::trace!("Ignoring write_buffer of size 0");
             return Ok(());
         };
@@ -791,6 +793,8 @@ impl Queue {
 
         let dst_raw = dst.try_raw(&snatch_guard)?;
 
+        // This must happen after parameter validation (so that errors are reported
+        // as required by the spec), but before any side effects.
         if size.width == 0 || size.height == 0 || size.depth_or_array_layers == 0 {
             log::trace!("Ignoring write_texture of size 0");
             return Ok(());
@@ -963,11 +967,6 @@ impl Queue {
 
         self.device.check_is_valid()?;
 
-        if size.width == 0 || size.height == 0 || size.depth_or_array_layers == 0 {
-            log::trace!("Ignoring write_texture of size 0");
-            return Ok(());
-        }
-
         let mut needs_flag = false;
         needs_flag |= matches!(source.source, wgt::ExternalImageSource::OffscreenCanvas(_));
         needs_flag |= source.origin != wgt::Origin2d::ZERO;
@@ -1051,6 +1050,13 @@ impl Queue {
 
         let (selector, dst_base) = extract_texture_selector(&destination, &size, &dst)?;
 
+        // This must happen after parameter validation (so that errors are reported
+        // as required by the spec), but before any side effects.
+        if size.width == 0 || size.height == 0 || size.depth_or_array_layers == 0 {
+            log::trace!("Ignoring copy_external_image_to_texture of size 0");
+            return Ok(());
+        }
+
         let mut pending_writes = self.pending_writes.lock();
         let encoder = pending_writes.activate();
 

wgpu-core/src/device/resource.rs

@@ -1424,6 +1424,11 @@ impl Device {
             .map_err(|error| CreateTextureError::MissingFeatures(desc.format, error))?;
 
         if desc.sample_count > 1 {
+            // <https://www.w3.org/TR/2025/CRD-webgpu-20251120/#:~:text=If%20descriptor%2EsampleCount%20%3E%201>
+            //
+            // Note that there are also some checks related to the sample count
+            // in [`conv::check_texture_dimension_size`].
+
             if desc.mip_level_count != 1 {
                 return Err(CreateTextureError::InvalidMipLevelCount {
                     requested: desc.mip_level_count,