Bug 1922854: Update wgpu to ee0d1703 (2024-10-04). r=webgpu-reviewers,supply-chain-reviewers,nical,ErichDonGubler

Differential Revision: https://phabricator.services.mozilla.com/D224614

.cargo/config.toml.in

@@ -25,9 +25,9 @@ git = "https://github.com/franziskuskiefer/cose-rust"
 rev = "43c22248d136c8b38fe42ea709d08da6355cf04b"
 replace-with = "vendored-sources"
 
-[source."git+https://github.com/gfx-rs/wgpu?rev=3fda684eb9e69c78b16312a3e927e3ea82e853d1"]
+[source."git+https://github.com/gfx-rs/wgpu?rev=ee0d1703e5f4a267ce9b87d50b824190b45b5a79"]
 git = "https://github.com/gfx-rs/wgpu"
-rev = "3fda684eb9e69c78b16312a3e927e3ea82e853d1"
+rev = "ee0d1703e5f4a267ce9b87d50b824190b45b5a79"
 replace-with = "vendored-sources"
 
 [source."git+https://github.com/hsivonen/any_all_workaround?rev=7fb1b7034c9f172aade21ee1c8554e8d8a48af80"]

Cargo.lock

@@ -4216,7 +4216,7 @@ checksum = "a2983372caf4480544083767bf2d27defafe32af49ab4df3a0b7fc90793a3664"
 [[package]]
 name = "naga"
 version = "22.0.0"
-source = "git+https://github.com/gfx-rs/wgpu?rev=3fda684eb9e69c78b16312a3e927e3ea82e853d1#3fda684eb9e69c78b16312a3e927e3ea82e853d1"
+source = "git+https://github.com/gfx-rs/wgpu?rev=ee0d1703e5f4a267ce9b87d50b824190b45b5a79#ee0d1703e5f4a267ce9b87d50b824190b45b5a79"
 dependencies = [
  "arrayvec",
  "bit-set",
@@ -6078,18 +6078,18 @@ checksum = "aac81b6fd6beb5884b0cf3321b8117e6e5d47ecb6fc89f414cfdcca8b2fe2dd8"
 
 [[package]]
 name = "thiserror"
-version = "1.0.63"
+version = "1.0.64"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "c0342370b38b6a11b6cc11d6a805569958d54cfa061a29969c3b5ce2ea405724"
+checksum = "d50af8abc119fb8bb6dbabcfa89656f46f84aa0ac7688088608076ad2b459a84"
 dependencies = [
  "thiserror-impl",
 ]
 
 [[package]]
 name = "thiserror-impl"
-version = "1.0.63"
+version = "1.0.64"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "a4558b58466b9ad7ca0f102865eccc95938dca1a74a856f2b57b6629050da261"
+checksum = "08904e7672f5eb876eaaf87e0ce17857500934f4981c4a0ab2b4aa98baac7fc3"
 dependencies = [
  "proc-macro2",
  "quote",
@@ -6410,9 +6410,9 @@ checksum = "c0edd1e5b14653f783770bce4a4dabb4a5108a5370a5f5d8cfe8710c361f6c8b"
 
 [[package]]
 name = "unicode-xid"
-version = "0.2.5"
+version = "0.2.6"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "229730647fbc343e3a80e463c1db7f78f3855d3f3739bee0dda773c9a037c90a"
+checksum = "ebc1c04c71510c7f702b52b7c350734c9ff1295c464a03335b00bb84fc54f853"
 
 [[package]]
 name = "uniffi"
@@ -7013,7 +7013,7 @@ dependencies = [
 [[package]]
 name = "wgpu-core"
 version = "22.0.0"
-source = "git+https://github.com/gfx-rs/wgpu?rev=3fda684eb9e69c78b16312a3e927e3ea82e853d1#3fda684eb9e69c78b16312a3e927e3ea82e853d1"
+source = "git+https://github.com/gfx-rs/wgpu?rev=ee0d1703e5f4a267ce9b87d50b824190b45b5a79#ee0d1703e5f4a267ce9b87d50b824190b45b5a79"
 dependencies = [
  "arrayvec",
  "bit-vec",
@@ -7038,7 +7038,7 @@ dependencies = [
 [[package]]
 name = "wgpu-hal"
 version = "22.0.0"
-source = "git+https://github.com/gfx-rs/wgpu?rev=3fda684eb9e69c78b16312a3e927e3ea82e853d1#3fda684eb9e69c78b16312a3e927e3ea82e853d1"
+source = "git+https://github.com/gfx-rs/wgpu?rev=ee0d1703e5f4a267ce9b87d50b824190b45b5a79#ee0d1703e5f4a267ce9b87d50b824190b45b5a79"
 dependencies = [
  "android_system_properties",
  "arrayvec",
@@ -7077,7 +7077,7 @@ dependencies = [
 [[package]]
 name = "wgpu-types"
 version = "22.0.0"
-source = "git+https://github.com/gfx-rs/wgpu?rev=3fda684eb9e69c78b16312a3e927e3ea82e853d1#3fda684eb9e69c78b16312a3e927e3ea82e853d1"
+source = "git+https://github.com/gfx-rs/wgpu?rev=ee0d1703e5f4a267ce9b87d50b824190b45b5a79#ee0d1703e5f4a267ce9b87d50b824190b45b5a79"
 dependencies = [
  "bitflags 2.6.0",
  "js-sys",

gfx/wgpu_bindings/Cargo.toml

@@ -17,7 +17,7 @@ default = []
 [dependencies.wgc]
 package = "wgpu-core"
 git = "https://github.com/gfx-rs/wgpu"
-rev = "3fda684eb9e69c78b16312a3e927e3ea82e853d1"
+rev = "ee0d1703e5f4a267ce9b87d50b824190b45b5a79"
 # TODO: remove the replay feature on the next update containing https://github.com/gfx-rs/wgpu/pull/5182
 features = ["serde", "replay", "trace", "strict_asserts", "wgsl", "api_log_info"]
 
@@ -26,32 +26,32 @@ features = ["serde", "replay", "trace", "strict_asserts", "wgsl", "api_log_info"
 [target.'cfg(any(target_os = "macos", target_os = "ios"))'.dependencies.wgc]
 package = "wgpu-core"
 git = "https://github.com/gfx-rs/wgpu"
-rev = "3fda684eb9e69c78b16312a3e927e3ea82e853d1"
+rev = "ee0d1703e5f4a267ce9b87d50b824190b45b5a79"
 features = ["metal"]
 
 # We want the wgpu-core Direct3D backends on Windows.
 [target.'cfg(windows)'.dependencies.wgc]
 package = "wgpu-core"
 git = "https://github.com/gfx-rs/wgpu"
-rev = "3fda684eb9e69c78b16312a3e927e3ea82e853d1"
+rev = "ee0d1703e5f4a267ce9b87d50b824190b45b5a79"
 features = ["dx12"]
 
 # We want the wgpu-core Vulkan backend on Linux and Windows.
 [target.'cfg(any(windows, all(unix, not(any(target_os = "macos", target_os = "ios")))))'.dependencies.wgc]
 package = "wgpu-core"
 git = "https://github.com/gfx-rs/wgpu"
-rev = "3fda684eb9e69c78b16312a3e927e3ea82e853d1"
+rev = "ee0d1703e5f4a267ce9b87d50b824190b45b5a79"
 features = ["vulkan"]
 
 [dependencies.wgt]
 package = "wgpu-types"
 git = "https://github.com/gfx-rs/wgpu"
-rev = "3fda684eb9e69c78b16312a3e927e3ea82e853d1"
+rev = "ee0d1703e5f4a267ce9b87d50b824190b45b5a79"
 
 [dependencies.wgh]
 package = "wgpu-hal"
 git = "https://github.com/gfx-rs/wgpu"
-rev = "3fda684eb9e69c78b16312a3e927e3ea82e853d1"
+rev = "ee0d1703e5f4a267ce9b87d50b824190b45b5a79"
 features = ["oom_panic", "device_lost_panic", "internal_error_panic"]
 
 [target.'cfg(windows)'.dependencies]

gfx/wgpu_bindings/moz.yaml

@@ -20,11 +20,11 @@ origin:
 
   # Human-readable identifier for this version/release
   # Generally "version NNN", "tag SSS", "bookmark SSS"
-  release: 3fda684eb9e69c78b16312a3e927e3ea82e853d1 (2024-09-18T15:01:51Z).
+  release: ee0d1703e5f4a267ce9b87d50b824190b45b5a79 (Fri Oct 4 13:21:59 2024 -0400).
 
   # Revision to pull in
   # Must be a long or short commit SHA (long preferred)
-  revision: 3fda684eb9e69c78b16312a3e927e3ea82e853d1
+  revision: ee0d1703e5f4a267ce9b87d50b824190b45b5a79
 
   license: ['MIT', 'Apache-2.0']
 

gfx/wgpu_bindings/src/client.rs

@@ -10,8 +10,8 @@ use crate::{
 
 use crate::SwapChainId;
 
-use wgc::{id, identity::IdentityManager};
-use wgt::TextureFormat;
+use wgc::{command::RenderBundleEncoder, id, identity::IdentityManager};
+use wgt::{BufferAddress, BufferSize, DynamicOffset, IndexFormat, TextureFormat};
 
 use wgc::id::markers;
 
@@ -1287,3 +1287,146 @@ pub extern "C" fn wgpu_client_use_external_texture_in_swapChain(
 
     supported
 }
+
+#[no_mangle]
+pub unsafe extern "C" fn wgpu_render_bundle_set_bind_group(
+    bundle: &mut RenderBundleEncoder,
+    index: u32,
+    bind_group_id: Option<id::BindGroupId>,
+    offsets: *const DynamicOffset,
+    offset_length: usize,
+) {
+    wgc::command::bundle_ffi::wgpu_render_bundle_set_bind_group(
+        bundle,
+        index,
+        bind_group_id,
+        offsets,
+        offset_length,
+    )
+}
+
+#[no_mangle]
+pub extern "C" fn wgpu_render_bundle_set_pipeline(
+    bundle: &mut RenderBundleEncoder,
+    pipeline_id: id::RenderPipelineId,
+) {
+    wgc::command::bundle_ffi::wgpu_render_bundle_set_pipeline(bundle, pipeline_id)
+}
+
+#[no_mangle]
+pub extern "C" fn wgpu_render_bundle_set_vertex_buffer(
+    bundle: &mut RenderBundleEncoder,
+    slot: u32,
+    buffer_id: id::BufferId,
+    offset: BufferAddress,
+    size: Option<BufferSize>,
+) {
+    wgc::command::bundle_ffi::wgpu_render_bundle_set_vertex_buffer(
+        bundle, slot, buffer_id, offset, size,
+    )
+}
+
+#[no_mangle]
+pub extern "C" fn wgpu_render_bundle_set_index_buffer(
+    encoder: &mut RenderBundleEncoder,
+    buffer: id::BufferId,
+    index_format: IndexFormat,
+    offset: BufferAddress,
+    size: Option<BufferSize>,
+) {
+    wgc::command::bundle_ffi::wgpu_render_bundle_set_index_buffer(
+        encoder,
+        buffer,
+        index_format,
+        offset,
+        size,
+    )
+}
+
+#[no_mangle]
+pub unsafe extern "C" fn wgpu_render_bundle_set_push_constants(
+    pass: &mut RenderBundleEncoder,
+    stages: wgt::ShaderStages,
+    offset: u32,
+    size_bytes: u32,
+    data: *const u8,
+) {
+    wgc::command::bundle_ffi::wgpu_render_bundle_set_push_constants(
+        pass, stages, offset, size_bytes, data,
+    )
+}
+
+#[no_mangle]
+pub extern "C" fn wgpu_render_bundle_draw(
+    bundle: &mut RenderBundleEncoder,
+    vertex_count: u32,
+    instance_count: u32,
+    first_vertex: u32,
+    first_instance: u32,
+) {
+    wgc::command::bundle_ffi::wgpu_render_bundle_draw(
+        bundle,
+        vertex_count,
+        instance_count,
+        first_vertex,
+        first_instance,
+    )
+}
+
+#[no_mangle]
+pub extern "C" fn wgpu_render_bundle_draw_indexed(
+    bundle: &mut RenderBundleEncoder,
+    index_count: u32,
+    instance_count: u32,
+    first_index: u32,
+    base_vertex: i32,
+    first_instance: u32,
+) {
+    wgc::command::bundle_ffi::wgpu_render_bundle_draw_indexed(
+        bundle,
+        index_count,
+        instance_count,
+        first_index,
+        base_vertex,
+        first_instance,
+    )
+}
+
+#[no_mangle]
+pub extern "C" fn wgpu_render_bundle_draw_indirect(
+    bundle: &mut RenderBundleEncoder,
+    buffer_id: id::BufferId,
+    offset: BufferAddress,
+) {
+    wgc::command::bundle_ffi::wgpu_render_bundle_draw_indirect(bundle, buffer_id, offset)
+}
+
+#[no_mangle]
+pub extern "C" fn wgpu_render_bundle_draw_indexed_indirect(
+    bundle: &mut RenderBundleEncoder,
+    buffer_id: id::BufferId,
+    offset: BufferAddress,
+) {
+    wgc::command::bundle_ffi::wgpu_render_bundle_draw_indexed_indirect(bundle, buffer_id, offset)
+}
+
+#[no_mangle]
+pub unsafe extern "C" fn wgpu_render_bundle_push_debug_group(
+    _bundle: &mut RenderBundleEncoder,
+    _label: RawString,
+) {
+    wgc::command::bundle_ffi::wgpu_render_bundle_push_debug_group(_bundle, _label)
+}
+
+#[no_mangle]
+pub extern "C" fn wgpu_render_bundle_pop_debug_group(_bundle: &mut RenderBundleEncoder) {
+    wgc::command::bundle_ffi::wgpu_render_bundle_pop_debug_group(_bundle)
+}
+
+#[no_mangle]
+pub unsafe extern "C" fn wgpu_render_bundle_insert_debug_marker(
+    _bundle: &mut RenderBundleEncoder,
+    _label: RawString,
+) {
+    wgc::command::bundle_ffi::wgpu_render_bundle_insert_debug_marker(_bundle, _label)
+}

gfx/wgpu_bindings/src/server.rs

@@ -1302,7 +1302,7 @@ pub unsafe extern "C" fn wgpu_server_queue_submit(
     let result = global.queue_submit(self_id, command_buffers);
 
     match result {
-        Err(err) => {
+        Err((_index, err)) => {
             error_buf.init(err);
             return 0;
         }

supply-chain/audits.toml

@@ -3237,12 +3237,12 @@ delta = "0.20.0 -> 22.0.0"
 
 [[audits.naga]]
 who = [
-    "Jim Blandy <jimb@red-bean.com>",
     "Teodor Tanasoaia <ttanasoaia@mozilla.com>",
     "Erich Gubler <erichdongubler@gmail.com>",
+    "Jim Blandy <jimb@red-bean.com>",
 ]
 criteria = "safe-to-deploy"
-delta = "22.0.0 -> 22.0.0@git:3fda684eb9e69c78b16312a3e927e3ea82e853d1"
+delta = "22.0.0 -> 22.0.0@git:ee0d1703e5f4a267ce9b87d50b824190b45b5a79"
 importable = false
 
 [[audits.net2]]
@@ -4848,6 +4848,11 @@ who = "Teodor Tanasoaia <ttanasoaia@mozilla.com>"
 criteria = "safe-to-deploy"
 delta = "0.2.4 -> 0.2.5"
 
+[[audits.unicode-xid]]
+who = "Jim Blandy <jimb@red-bean.com>"
+criteria = "safe-to-deploy"
+delta = "0.2.5 -> 0.2.6"
+
 [[audits.uniffi]]
 who = "Travis Long <tlong@mozilla.com>"
 criteria = "safe-to-deploy"
@@ -5275,12 +5280,12 @@ delta = "0.20.0 -> 22.0.0"
 
 [[audits.wgpu-core]]
 who = [
-    "Jim Blandy <jimb@red-bean.com>",
     "Teodor Tanasoaia <ttanasoaia@mozilla.com>",
     "Erich Gubler <erichdongubler@gmail.com>",
+    "Jim Blandy <jimb@red-bean.com>",
 ]
 criteria = "safe-to-deploy"
-delta = "22.0.0 -> 22.0.0@git:3fda684eb9e69c78b16312a3e927e3ea82e853d1"
+delta = "22.0.0 -> 22.0.0@git:ee0d1703e5f4a267ce9b87d50b824190b45b5a79"
 importable = false
 
 [[audits.wgpu-hal]]
@@ -5348,12 +5353,12 @@ delta = "0.20.0 -> 22.0.0"
 
 [[audits.wgpu-hal]]
 who = [
-    "Jim Blandy <jimb@red-bean.com>",
     "Teodor Tanasoaia <ttanasoaia@mozilla.com>",
     "Erich Gubler <erichdongubler@gmail.com>",
+    "Jim Blandy <jimb@red-bean.com>",
 ]
 criteria = "safe-to-deploy"
-delta = "22.0.0 -> 22.0.0@git:3fda684eb9e69c78b16312a3e927e3ea82e853d1"
+delta = "22.0.0 -> 22.0.0@git:ee0d1703e5f4a267ce9b87d50b824190b45b5a79"
 importable = false
 
 [[audits.wgpu-types]]
@@ -5421,12 +5426,12 @@ delta = "0.20.0 -> 22.0.0"
 
 [[audits.wgpu-types]]
 who = [
-    "Jim Blandy <jimb@red-bean.com>",
     "Teodor Tanasoaia <ttanasoaia@mozilla.com>",
     "Erich Gubler <erichdongubler@gmail.com>",
+    "Jim Blandy <jimb@red-bean.com>",
 ]
 criteria = "safe-to-deploy"
-delta = "22.0.0 -> 22.0.0@git:3fda684eb9e69c78b16312a3e927e3ea82e853d1"
+delta = "22.0.0 -> 22.0.0@git:ee0d1703e5f4a267ce9b87d50b824190b45b5a79"
 importable = false
 
 [[audits.whatsys]]

supply-chain/imports.lock

@@ -687,15 +687,15 @@ user-login = "BurntSushi"
 user-name = "Andrew Gallant"
 
 [[publisher.thiserror]]
-version = "1.0.63"
-when = "2024-07-17"
+version = "1.0.64"
+when = "2024-09-22"
 user-id = 3618
 user-login = "dtolnay"
 user-name = "David Tolnay"
 
 [[publisher.thiserror-impl]]
-version = "1.0.63"
-when = "2024-07-17"
+version = "1.0.64"
+when = "2024-09-22"
 user-id = 3618
 user-login = "dtolnay"
 user-name = "David Tolnay"

testing/web-platform/mozilla/meta/webgpu/cts/webgpu/compat/api/validation/render_pipeline/unsupported_wgsl/cts.https.html.ini

@@ -183,24 +183,30 @@
 
 [cts.https.html?q=webgpu:compat,api,validation,render_pipeline,unsupported_wgsl:unsupportedStorageTextureFormats,computePipeline:*]
   [:format="rg32float";entryPoint="csWithStorageUsage";async=false]
+    expected: FAIL
 
   [:format="rg32float";entryPoint="csWithStorageUsage";async=true]
+    expected: FAIL
 
   [:format="rg32float";entryPoint="csWithoutStorageUsage";async=false]
 
   [:format="rg32float";entryPoint="csWithoutStorageUsage";async=true]
 
   [:format="rg32sint";entryPoint="csWithStorageUsage";async=false]
+    expected: FAIL
 
   [:format="rg32sint";entryPoint="csWithStorageUsage";async=true]
+    expected: FAIL
 
   [:format="rg32sint";entryPoint="csWithoutStorageUsage";async=false]
 
   [:format="rg32sint";entryPoint="csWithoutStorageUsage";async=true]
 
   [:format="rg32uint";entryPoint="csWithStorageUsage";async=false]
+    expected: FAIL
 
   [:format="rg32uint";entryPoint="csWithStorageUsage";async=true]
+    expected: FAIL
 
   [:format="rg32uint";entryPoint="csWithoutStorageUsage";async=false]
 

third_party/rust/naga/Cargo.toml

@@ -96,10 +96,10 @@ optional = true
 version = "1.4.1"
 
 [dependencies.thiserror]
-version = "1.0.63"
+version = "1.0.64"
 
 [dependencies.unicode-xid]
-version = "0.2.5"
+version = "0.2.6"
 optional = true
 
 [dev-dependencies]

third_party/rust/naga/src/back/dot/mod.rs

@@ -698,7 +698,7 @@ fn write_function_expressions(
             E::RayQueryGetIntersection { query, committed } => {
                 edges.insert("", query);
                 let ty = if committed { "Committed" } else { "Candidate" };
-                (format!("rayQueryGet{}Intersection", ty).into(), 4)
+                (format!("rayQueryGet{ty}Intersection").into(), 4)
             }
             E::SubgroupBallotResult => ("SubgroupBallotResult".into(), 4),
             E::SubgroupOperationResult { .. } => ("SubgroupOperationResult".into(), 4),

third_party/rust/naga/src/back/glsl/mod.rs

@@ -2645,15 +2645,15 @@ impl<'a, W: Write> Writer<'a, W> {
                 match literal {
                     // Floats are written using `Debug` instead of `Display` because it always appends the
                     // decimal part even it's zero which is needed for a valid glsl float constant
-                    crate::Literal::F64(value) => write!(self.out, "{:?}LF", value)?,
-                    crate::Literal::F32(value) => write!(self.out, "{:?}", value)?,
+                    crate::Literal::F64(value) => write!(self.out, "{value:?}LF")?,
+                    crate::Literal::F32(value) => write!(self.out, "{value:?}")?,
                     // Unsigned integers need a `u` at the end
                     //
                     // While `core` doesn't necessarily need it, it's allowed and since `es` needs it we
                     // always write it as the extra branch wouldn't have any benefit in readability
-                    crate::Literal::U32(value) => write!(self.out, "{}u", value)?,
-                    crate::Literal::I32(value) => write!(self.out, "{}", value)?,
-                    crate::Literal::Bool(value) => write!(self.out, "{}", value)?,
+                    crate::Literal::U32(value) => write!(self.out, "{value}u")?,
+                    crate::Literal::I32(value) => write!(self.out, "{value}")?,
+                    crate::Literal::Bool(value) => write!(self.out, "{value}")?,
                     crate::Literal::I64(_) => {
                         return Err(Error::Custom("GLSL has no 64-bit integer type".into()));
                     }
@@ -4614,7 +4614,7 @@ impl<'a, W: Write> Writer<'a, W> {
 
                 for i in 0..count.get() {
                     // Add the array accessor and recurse.
-                    segments.push(format!("[{}]", i));
+                    segments.push(format!("[{i}]"));
                     self.collect_push_constant_items(base, segments, layouter, offset, items);
                     segments.pop();
                 }

third_party/rust/naga/src/back/hlsl/help.rs

@@ -1046,8 +1046,7 @@ impl<'a, W: Write> super::Writer<'a, W> {
                         }
                         ref other => {
                             return Err(super::Error::Unimplemented(format!(
-                                "Array length of base {:?}",
-                                other
+                                "Array length of base {other:?}"
                             )))
                         }
                     };

third_party/rust/naga/src/back/hlsl/storage.rs

@@ -350,7 +350,7 @@ impl<W: fmt::Write> super::Writer<'_, W> {
                     self.write_store_value(module, &value, func_ctx)?;
                     writeln!(self.out, "));")?;
                 } else {
-                    write!(self.out, "{}{}.Store(", level, var_name)?;
+                    write!(self.out, "{level}{var_name}.Store(")?;
                     self.write_storage_address(module, &chain, func_ctx)?;
                     write!(self.out, ", ")?;
                     self.write_store_value(module, &value, func_ctx)?;

third_party/rust/naga/src/back/hlsl/writer.rs

@@ -965,7 +965,7 @@ impl<'a, W: fmt::Write> super::Writer<'a, W> {
         let constant = &module.constants[handle];
         self.write_type(module, constant.ty)?;
         let name = &self.names[&NameKey::Constant(handle)];
-        write!(self.out, " {}", name)?;
+        write!(self.out, " {name}")?;
         // Write size for array type
         if let TypeInner::Array { base, size, .. } = module.types[constant.ty].inner {
             self.write_array_size(module, base, size)?;
@@ -2383,11 +2383,11 @@ impl<'a, W: fmt::Write> super::Writer<'a, W> {
                 // decimal part even it's zero
                 crate::Literal::F64(value) => write!(self.out, "{value:?}L")?,
                 crate::Literal::F32(value) => write!(self.out, "{value:?}")?,
-                crate::Literal::U32(value) => write!(self.out, "{}u", value)?,
-                crate::Literal::I32(value) => write!(self.out, "{}", value)?,
-                crate::Literal::U64(value) => write!(self.out, "{}uL", value)?,
-                crate::Literal::I64(value) => write!(self.out, "{}L", value)?,
-                crate::Literal::Bool(value) => write!(self.out, "{}", value)?,
+                crate::Literal::U32(value) => write!(self.out, "{value}u")?,
+                crate::Literal::I32(value) => write!(self.out, "{value}")?,
+                crate::Literal::U64(value) => write!(self.out, "{value}uL")?,
+                crate::Literal::I64(value) => write!(self.out, "{value}L")?,
+                crate::Literal::Bool(value) => write!(self.out, "{value}")?,
                 crate::Literal::AbstractInt(_) | crate::Literal::AbstractFloat(_) => {
                     return Err(Error::Custom(
                         "Abstract types should not appear in IR presented to backends".into(),

third_party/rust/naga/src/back/msl/mod.rs

@@ -437,8 +437,7 @@ impl Options {
                     })
                 }
                 LocationMode::Uniform => Err(Error::GenericValidation(format!(
-                    "Unexpected Binding::Location({}) for the Uniform mode",
-                    location
+                    "Unexpected Binding::Location({location}) for the Uniform mode"
                 ))),
             },
         }

third_party/rust/naga/src/back/msl/writer.rs

@@ -782,6 +782,9 @@ impl<W: Write> Writer<W> {
     ///
     /// To make our output a bit more legible, we pull the condition out into a
     /// preprocessor macro defined at the top of the module.
+    ///
+    /// This approach is also used by Chromium WebGPU's Dawn shader compiler, as of
+    /// <https://github.com/google/dawn/commit/ffd485c685040edb1e678165dcbf0e841cfa0298>.
     fn emit_loop_reachable_macro(&mut self) -> BackendResult {
         if !self.loop_reachable_macro_name.is_empty() {
             return Ok(());
@@ -3817,12 +3820,11 @@ impl<W: Write> Writer<W> {
                     writeln!(self.out)?;
                     writeln!(
                         self.out,
-                        "{} {defined_func_name}({arg_type_name} arg) {{
+                        "{struct_name} {defined_func_name}({arg_type_name} arg) {{
     {other_type_name} other;
     {arg_type_name} fract = {NAMESPACE}::{called_func_name}(arg, other);
-    return {}{{ fract, other }};
-}}",
-                        struct_name, struct_name
+    return {struct_name}{{ fract, other }};
+}}"
                     )?;
                 }
                 &crate::PredeclaredType::AtomicCompareExchangeWeakResult { .. } => {}

third_party/rust/naga/src/back/pipeline_constants.rs

@@ -14,7 +14,10 @@ use thiserror::Error;
 pub enum PipelineConstantError {
     #[error("Missing value for pipeline-overridable constant with identifier string: '{0}'")]
     MissingValue(String),
-    #[error("Source f64 value needs to be finite (NaNs and Inifinites are not allowed) for number destinations")]
+    #[error(
+        "Source f64 value needs to be finite ({}) for number destinations",
+        "NaNs and Inifinites are not allowed"
+    )]
     SrcNeedsToBeFinite,
     #[error("Source f64 value doesn't fit in destination")]
     DstRangeTooSmall,

third_party/rust/naga/src/back/spv/block.rs

@@ -4,8 +4,7 @@ Implementations for `BlockContext` methods.
 
 use super::{
     helpers, index::BoundsCheckResult, make_local, selection::Selection, Block, BlockContext,
-    Dimension, Error, Instruction, LocalType, LookupType, LoopContext, ResultMember, Writer,
-    WriterFlags,
+    Dimension, Error, Instruction, LocalType, LookupType, ResultMember, Writer, WriterFlags,
 };
 use crate::{arena::Handle, proc::TypeResolution, Statement};
 use spirv::Word;
@@ -39,7 +38,7 @@ enum ExpressionPointer {
 }
 
 /// The termination statement to be added to the end of the block
-pub enum BlockExit {
+enum BlockExit {
     /// Generates an OpReturn (void return)
     Return,
     /// Generates an OpBranch to the specified block
@@ -60,6 +59,36 @@ pub enum BlockExit {
     },
 }
 
+/// What code generation did with a provided [`BlockExit`] value.
+///
+/// A function that accepts a [`BlockExit`] argument should return a value of
+/// this type, to indicate whether the code it generated ended up using the
+/// provided exit, or ignored it and did a non-local exit of some other kind
+/// (say, [`Break`] or [`Continue`]). Some callers must use this information to
+/// decide whether to generate the target block at all.
+///
+/// [`Break`]: Statement::Break
+/// [`Continue`]: Statement::Continue
+#[must_use]
+enum BlockExitDisposition {
+    /// The generated code used the provided `BlockExit` value. If it included a
+    /// block label, the caller should be sure to actually emit the block it
+    /// refers to.
+    Used,
+
+    /// The generated code did not use the provided `BlockExit` value. If it
+    /// included a block label, the caller should not bother to actually emit
+    /// the block it refers to, unless it knows the block is needed for
+    /// something else.
+    Discarded,
+}
+
+#[derive(Clone, Copy, Default)]
+struct LoopContext {
+    continuing_id: Option<Word>,
+    break_id: Option<Word>,
+}
+
 #[derive(Debug)]
 pub(crate) struct DebugInfoInner<'a> {
     pub source_code: &'a str,
@@ -200,10 +229,7 @@ impl<'w> BlockContext<'w> {
     fn is_intermediate(&self, expr_handle: Handle<crate::Expression>) -> bool {
         match self.ir_function.expressions[expr_handle] {
             crate::Expression::GlobalVariable(handle) => {
-                match self.ir_module.global_variables[handle].space {
-                    crate::AddressSpace::Handle => false,
-                    _ => true,
-                }
+                self.ir_module.global_variables[handle].space != crate::AddressSpace::Handle
             }
             crate::Expression::LocalVariable(_) => true,
             crate::Expression::FunctionArgument(index) => {
@@ -346,6 +372,32 @@ impl<'w> BlockContext<'w> {
 
                         load_id
                     }
+                    crate::TypeInner::Array {
+                        base: ty_element, ..
+                    } => {
+                        let index_id = self.cached[index];
+                        let base_id = self.cached[base];
+                        let base_ty = match self.fun_info[base].ty {
+                            TypeResolution::Handle(handle) => handle,
+                            TypeResolution::Value(_) => {
+                                return Err(Error::Validation(
+                                    "Array types should always be in the arena",
+                                ))
+                            }
+                        };
+                        let (id, variable) = self.writer.promote_access_expression_to_variable(
+                            &self.ir_module.types,
+                            result_type_id,
+                            base_id,
+                            base_ty,
+                            index_id,
+                            ty_element,
+                            block,
+                        )?;
+                        self.function.internal_variables.push(variable);
+                        id
+                    }
+                    // wgpu#4337: Support `crate::TypeInner::Matrix`
                     ref other => {
                         log::error!(
                             "Unable to access base {:?} of type {:?}",
@@ -353,7 +405,7 @@ impl<'w> BlockContext<'w> {
                             other
                         );
                         return Err(Error::Validation(
-                            "only vectors may be dynamically indexed by value",
+                            "only vectors and arrays may be dynamically indexed by value",
                         ));
                     }
                 }
@@ -2037,14 +2089,30 @@ impl<'w> BlockContext<'w> {
         }
     }
 
-    pub(super) fn write_block(
+    /// Generate one or more SPIR-V blocks for `naga_block`.
+    ///
+    /// Use `label_id` as the label for the SPIR-V entry point block.
+    ///
+    /// If control reaches the end of the SPIR-V block, terminate it according
+    /// to `exit`. This function's return value indicates whether it acted on
+    /// this parameter or not; see [`BlockExitDisposition`].
+    ///
+    /// If the block contains [`Break`] or [`Continue`] statements,
+    /// `loop_context` supplies the labels of the SPIR-V blocks to jump to. If
+    /// either of these labels are `None`, then it should have been a Naga
+    /// validation error for the corresponding statement to occur in this
+    /// context.
+    ///
+    /// [`Break`]: Statement::Break
+    /// [`Continue`]: Statement::Continue
+    fn write_block(
         &mut self,
         label_id: Word,
         naga_block: &crate::Block,
         exit: BlockExit,
         loop_context: LoopContext,
         debug_info: Option<&DebugInfoInner>,
-    ) -> Result<(), Error> {
+    ) -> Result<BlockExitDisposition, Error> {
         let mut block = Block::new(label_id);
         for (statement, span) in naga_block.span_iter() {
             if let (Some(debug_info), false) = (
@@ -2080,7 +2148,7 @@ impl<'w> BlockContext<'w> {
                     self.function.consume(block, Instruction::branch(scope_id));
 
                     let merge_id = self.gen_id();
-                    self.write_block(
+                    let merge_used = self.write_block(
                         scope_id,
                         block_statements,
                         BlockExit::Branch { target: merge_id },
@@ -2088,7 +2156,14 @@ impl<'w> BlockContext<'w> {
                         debug_info,
                     )?;
 
-                    block = Block::new(merge_id);
+                    match merge_used {
+                        BlockExitDisposition::Used => {
+                            block = Block::new(merge_id);
+                        }
+                        BlockExitDisposition::Discarded => {
+                            return Ok(BlockExitDisposition::Discarded);
+                        }
+                    }
                 }
                 Statement::If {
                     condition,
@@ -2124,7 +2199,11 @@ impl<'w> BlockContext<'w> {
                     );
 
                     if let Some(block_id) = accept_id {
-                        self.write_block(
+                        // We can ignore the `BlockExitDisposition` returned here because,
+                        // even if `merge_id` is not actually reachable, it is always
+                        // referred to by the `OpSelectionMerge` instruction we emitted
+                        // earlier.
+                        let _ = self.write_block(
                             block_id,
                             accept,
                             BlockExit::Branch { target: merge_id },
@@ -2133,7 +2212,11 @@ impl<'w> BlockContext<'w> {
                         )?;
                     }
                     if let Some(block_id) = reject_id {
-                        self.write_block(
+                        // We can ignore the `BlockExitDisposition` returned here because,
+                        // even if `merge_id` is not actually reachable, it is always
+                        // referred to by the `OpSelectionMerge` instruction we emitted
+                        // earlier.
+                        let _ = self.write_block(
                             block_id,
                             reject,
                             BlockExit::Branch { target: merge_id },
@@ -2211,7 +2294,15 @@ impl<'w> BlockContext<'w> {
                         } else {
                             merge_id
                         };
-                        self.write_block(
+                        // We can ignore the `BlockExitDisposition` returned here because
+                        // `case_finish_id` is always referred to by either:
+                        //
+                        // - the `OpSwitch`, if it's the next case's label for a
+                        //   fall-through, or
+                        //
+                        // - the `OpSelectionMerge`, if it's the switch's overall merge
+                        //   block because there's no fall-through.
+                        let _ = self.write_block(
                             *label_id,
                             &case.body,
                             BlockExit::Branch {
@@ -2257,7 +2348,10 @@ impl<'w> BlockContext<'w> {
                     ));
                     self.function.consume(block, Instruction::branch(body_id));
 
-                    self.write_block(
+                    // We can ignore the `BlockExitDisposition` returned here because,
+                    // even if `continuing_id` is not actually reachable, it is always
+                    // referred to by the `OpLoopMerge` instruction we emitted earlier.
+                    let _ = self.write_block(
                         body_id,
                         body,
                         BlockExit::Branch {
@@ -2280,7 +2374,10 @@ impl<'w> BlockContext<'w> {
                         },
                     };
 
-                    self.write_block(
+                    // We can ignore the `BlockExitDisposition` returned here because,
+                    // even if `merge_id` is not actually reachable, it is always referred
+                    // to by the `OpLoopMerge` instruction we emitted earlier.
+                    let _ = self.write_block(
                         continuing_id,
                         continuing,
                         exit,
@@ -2296,14 +2393,14 @@ impl<'w> BlockContext<'w> {
                 Statement::Break => {
                     self.function
                         .consume(block, Instruction::branch(loop_context.break_id.unwrap()));
-                    return Ok(());
+                    return Ok(BlockExitDisposition::Discarded);
                 }
                 Statement::Continue => {
                     self.function.consume(
                         block,
                         Instruction::branch(loop_context.continuing_id.unwrap()),
                     );
-                    return Ok(());
+                    return Ok(BlockExitDisposition::Discarded);
                 }
                 Statement::Return { value: Some(value) } => {
                     let value_id = self.cached[value];
@@ -2322,15 +2419,15 @@ impl<'w> BlockContext<'w> {
                         None => Instruction::return_value(value_id),
                     };
                     self.function.consume(block, instruction);
-                    return Ok(());
+                    return Ok(BlockExitDisposition::Discarded);
                 }
                 Statement::Return { value: None } => {
                     self.function.consume(block, Instruction::return_void());
-                    return Ok(());
+                    return Ok(BlockExitDisposition::Discarded);
                 }
                 Statement::Kill => {
                     self.function.consume(block, Instruction::kill());
-                    return Ok(());
+                    return Ok(BlockExitDisposition::Discarded);
                 }
                 Statement::Barrier(flags) => {
                     self.writer.write_barrier(flags, &mut block);
@@ -2696,6 +2793,24 @@ impl<'w> BlockContext<'w> {
         };
 
         self.function.consume(block, termination);
+        Ok(BlockExitDisposition::Used)
+    }
+
+    pub(super) fn write_function_body(
+        &mut self,
+        entry_id: Word,
+        debug_info: Option<&DebugInfoInner>,
+    ) -> Result<(), Error> {
+        // We can ignore the `BlockExitDisposition` returned here because
+        // `BlockExit::Return` doesn't refer to a block.
+        let _ = self.write_block(
+            entry_id,
+            &self.ir_function.body,
+            super::block::BlockExit::Return,
+            LoopContext::default(),
+            debug_info,
+        )?;
+
         Ok(())
     }
 }

third_party/rust/naga/src/back/spv/helpers.rs

@@ -85,20 +85,9 @@ impl crate::AddressSpace {
 
 /// Return true if the global requires a type decorated with `Block`.
 ///
-/// Vulkan spec v1.3 §15.6.2, "Descriptor Set Interface", says:
+/// See [`back::spv::GlobalVariable`] for details.
 ///
-/// > Variables identified with the `Uniform` storage class are used to
-/// > access transparent buffer backed resources. Such variables must
-/// > be:
-/// >
-/// > -   typed as `OpTypeStruct`, or an array of this type,
-/// >
-/// > -   identified with a `Block` or `BufferBlock` decoration, and
-/// >
-/// > -   laid out explicitly using the `Offset`, `ArrayStride`, and
-/// >     `MatrixStride` decorations as specified in §15.6.4, "Offset
-/// >     and Stride Assignment."
-// See `back::spv::GlobalVariable::access_id` for details.
+/// [`back::spv::GlobalVariable`]: super::GlobalVariable
 pub fn global_needs_wrapper(ir_module: &crate::Module, var: &crate::GlobalVariable) -> bool {
     match var.space {
         crate::AddressSpace::Uniform

third_party/rust/naga/src/back/spv/index.rs

@@ -11,16 +11,31 @@ use crate::{arena::Handle, proc::BoundsCheckPolicy};
 
 /// The results of performing a bounds check.
 ///
-/// On success, `write_bounds_check` returns a value of this type.
+/// On success, [`write_bounds_check`](BlockContext::write_bounds_check)
+/// returns a value of this type. The caller can assume that the right
+/// policy has been applied, and simply do what the variant says.
 pub(super) enum BoundsCheckResult {
     /// The index is statically known and in bounds, with the given value.
     KnownInBounds(u32),
 
     /// The given instruction computes the index to be used.
+    ///
+    /// When [`BoundsCheckPolicy::Restrict`] is in force, this is a
+    /// clamped version of the index the user supplied.
+    ///
+    /// When [`BoundsCheckPolicy::Unchecked`] is in force, this is
+    /// simply the index the user supplied. This variant indicates
+    /// that we couldn't prove statically that the index was in
+    /// bounds; otherwise we would have returned [`KnownInBounds`].
+    ///
+    /// [`KnownInBounds`]: BoundsCheckResult::KnownInBounds
     Computed(Word),
 
     /// The given instruction computes a boolean condition which is true
     /// if the index is in bounds.
+    ///
+    /// This is returned when [`BoundsCheckPolicy::ReadZeroSkipWrite`]
+    /// is in force.
     Conditional(Word),
 }
 
@@ -38,98 +53,163 @@ impl<'w> BlockContext<'w> {
     ///
     /// Given `array`, an expression referring a runtime-sized array, return the
     /// instruction id for the array's length.
+    ///
+    /// Runtime-sized arrays may only appear in the values of global
+    /// variables, which must have one of the following Naga types:
+    ///
+    /// 1. A runtime-sized array.
+    /// 2. A struct whose last member is a runtime-sized array.
+    /// 3. A binding array of 2.
+    ///
+    /// Thus, the expression `array` has the form of:
+    ///
+    /// - An optional [`AccessIndex`], for case 2, applied to...
+    /// - An optional [`Access`] or [`AccessIndex`], for case 3, applied to...
+    /// - A [`GlobalVariable`].
+    ///
+    /// The generated SPIR-V takes into account wrapped globals; see
+    /// [`back::spv::GlobalVariable`] for details.
+    ///
+    /// [`GlobalVariable`]: crate::Expression::GlobalVariable
+    /// [`AccessIndex`]: crate::Expression::AccessIndex
+    /// [`Access`]: crate::Expression::Access
+    /// [`base`]: crate::Expression::Access::base
+    /// [`back::spv::GlobalVariable`]: super::GlobalVariable
     pub(super) fn write_runtime_array_length(
         &mut self,
         array: Handle<crate::Expression>,
         block: &mut Block,
     ) -> Result<Word, Error> {
-        // Naga IR permits runtime-sized arrays as global variables, or as the
-        // final member of a struct that is a global variable, or one of these
-        // inside a buffer that is itself an element in a buffer bindings array.
-        // SPIR-V requires that runtime-sized arrays are wrapped in structs.
-        // See `helpers::global_needs_wrapper` and its uses.
-        let (opt_array_index_id, global_handle, opt_last_member_index) = match self
-            .ir_function
-            .expressions[array]
-        {
+        // The index into the binding array, if any.
+        let binding_array_index_id: Option<Word>;
+
+        // The handle to the Naga IR global we're referring to.
+        let global_handle: Handle<crate::GlobalVariable>;
+
+        // At the Naga type level, if the runtime-sized array is the final member of a
+        // struct, this is that member's index.
+        //
+        // This does not cover wrappers: if this backend wrapped the Naga global's
+        // type in a synthetic SPIR-V struct (see `global_needs_wrapper`), this is
+        // `None`.
+        let opt_last_member_index: Option<u32>;
+
+        // Inspect `array` and decide whether we have a binding array and/or an
+        // enclosing struct.
+        match self.ir_function.expressions[array] {
             crate::Expression::AccessIndex { base, index } => {
                 match self.ir_function.expressions[base] {
-                    // The global variable is an array of buffer bindings of structs,
-                    // we are accessing one of them with a static index,
-                    // and the last member of it.
                     crate::Expression::AccessIndex {
                         base: base_outer,
                         index: index_outer,
                     } => match self.ir_function.expressions[base_outer] {
+                        // An `AccessIndex` of an `AccessIndex` must be a
+                        // binding array holding structs whose last members are
+                        // runtime-sized arrays.
                         crate::Expression::GlobalVariable(handle) => {
                             let index_id = self.get_index_constant(index_outer);
-                            (Some(index_id), handle, Some(index))
+                            binding_array_index_id = Some(index_id);
+                            global_handle = handle;
+                            opt_last_member_index = Some(index);
+                        }
+                        _ => {
+                            return Err(Error::Validation(
+                                "array length expression: AccessIndex(AccessIndex(Global))",
+                            ))
                         }
-                        _ => return Err(Error::Validation("array length expression case-1a")),
                     },
-                    // The global variable is an array of buffer bindings of structs,
-                    // we are accessing one of them with a dynamic index,
-                    // and the last member of it.
                     crate::Expression::Access {
                         base: base_outer,
                         index: index_outer,
                     } => match self.ir_function.expressions[base_outer] {
+                        // Similarly, an `AccessIndex` of an `Access` must be a
+                        // binding array holding structs whose last members are
+                        // runtime-sized arrays.
                         crate::Expression::GlobalVariable(handle) => {
                             let index_id = self.cached[index_outer];
-                            (Some(index_id), handle, Some(index))
+                            binding_array_index_id = Some(index_id);
+                            global_handle = handle;
+                            opt_last_member_index = Some(index);
+                        }
+                        _ => {
+                            return Err(Error::Validation(
+                                "array length expression: AccessIndex(Access(Global))",
+                            ))
                         }
-                        _ => return Err(Error::Validation("array length expression case-1b")),
                     },
-                    // The global variable is a buffer, and we are accessing the last member.
                     crate::Expression::GlobalVariable(handle) => {
-                        let global = &self.ir_module.global_variables[handle];
-                        match self.ir_module.types[global.ty].inner {
-                            // The global variable is an array of buffer bindings of run-time arrays.
-                            crate::TypeInner::BindingArray { .. } => (Some(index), handle, None),
-                            // The global variable is a struct, and we are accessing the last member
-                            _ => (None, handle, Some(index)),
-                        }
+                        // An outer `AccessIndex` applied directly to a
+                        // `GlobalVariable`. Since binding arrays can only contain
+                        // structs, this must be referring to the last member of a
+                        // struct that is a runtime-sized array.
+                        binding_array_index_id = None;
+                        global_handle = handle;
+                        opt_last_member_index = Some(index);
+                    }
+                    _ => {
+                        return Err(Error::Validation(
+                            "array length expression: AccessIndex(<unexpected>)",
+                        ))
                     }
-                    _ => return Err(Error::Validation("array length expression case-1c")),
                 }
             }
-            // The global variable is an array of buffer bindings of arrays.
-            crate::Expression::Access { base, index } => match self.ir_function.expressions[base] {
-                crate::Expression::GlobalVariable(handle) => {
-                    let index_id = self.cached[index];
-                    let global = &self.ir_module.global_variables[handle];
-                    match self.ir_module.types[global.ty].inner {
-                        crate::TypeInner::BindingArray { .. } => (Some(index_id), handle, None),
-                        _ => return Err(Error::Validation("array length expression case-2a")),
-                    }
-                }
-                _ => return Err(Error::Validation("array length expression case-2b")),
-            },
-            // The global variable is a run-time array.
             crate::Expression::GlobalVariable(handle) => {
-                let global = &self.ir_module.global_variables[handle];
-                if !global_needs_wrapper(self.ir_module, global) {
-                    return Err(Error::Validation("array length expression case-3"));
-                }
-                (None, handle, None)
+                // A direct reference to a global variable. This must hold the
+                // runtime-sized array directly.
+                binding_array_index_id = None;
+                global_handle = handle;
+                opt_last_member_index = None;
             }
             _ => return Err(Error::Validation("array length expression case-4")),
         };
 
+        // The verifier should have checked this, but make sure the inspection above
+        // agrees with the type about whether a binding array is involved.
+        //
+        // Eventually we do want to support `binding_array<array<T>>`. This check
+        // ensures that whoever relaxes the validator will get an error message from
+        // us, not just bogus SPIR-V.
+        let global = &self.ir_module.global_variables[global_handle];
+        match (
+            &self.ir_module.types[global.ty].inner,
+            binding_array_index_id,
+        ) {
+            (&crate::TypeInner::BindingArray { .. }, Some(_)) => {}
+            (_, None) => {}
+            _ => {
+                return Err(Error::Validation(
+                    "array length expression: bad binding array inference",
+                ))
+            }
+        }
+
+        // SPIR-V allows runtime-sized arrays to appear only as the last member of a
+        // struct. Determine this member's index.
         let gvar = self.writer.global_variables[global_handle].clone();
         let global = &self.ir_module.global_variables[global_handle];
-        let (last_member_index, gvar_id) = match opt_last_member_index {
-            Some(index) => (index, gvar.access_id),
-            None => {
-                if !global_needs_wrapper(self.ir_module, global) {
-                    return Err(Error::Validation(
-                        "pointer to a global that is not a wrapped array",
-                    ));
-                }
+        let needs_wrapper = global_needs_wrapper(self.ir_module, global);
+        let (last_member_index, gvar_id) = match (opt_last_member_index, needs_wrapper) {
+            (Some(index), false) => {
+                // At the Naga type level, the runtime-sized array appears as the
+                // final member of a struct, whose index is `index`. We didn't need to
+                // wrap this, since the Naga type meets SPIR-V's requirements already.
+                (index, gvar.access_id)
+            }
+            (None, true) => {
+                // At the Naga type level, the runtime-sized array does not appear
+                // within a struct. We wrapped this in an OpTypeStruct with nothing
+                // else in it, so the index is zero. OpArrayLength wants the pointer
+                // to the wrapper struct, so use `gvar.var_id`.
                 (0, gvar.var_id)
             }
+            _ => {
+                return Err(Error::Validation(
+                    "array length expression: bad SPIR-V wrapper struct inference",
+                ));
+            }
         };
-        let structure_id = match opt_array_index_id {
+
+        let structure_id = match binding_array_index_id {
             // We are indexing inside a binding array, generate the access op.
             Some(index_id) => {
                 let element_type_id = match self.ir_module.types[global.ty].inner {
@@ -293,6 +373,8 @@ impl<'w> BlockContext<'w> {
 
     /// Write an index bounds comparison to `block`, if needed.
     ///
+    /// This is used to implement [`BoundsCheckPolicy::ReadZeroSkipWrite`].
+    ///
     /// If we're able to determine statically that `index` is in bounds for
     /// `sequence`, return `KnownInBounds(value)`, where `value` is the actual
     /// value of the index. (In principle, one could know that the index is in
@@ -413,11 +495,23 @@ impl<'w> BlockContext<'w> {
 
     /// Emit code for bounds checks for an array, vector, or matrix access.
     ///
-    /// This implements either `index_bounds_check_policy` or
-    /// `buffer_bounds_check_policy`, depending on the address space of the
-    /// pointer being accessed.
+    /// This tries to handle all the critical steps for bounds checks:
     ///
-    /// Return a `BoundsCheckResult` indicating how the index should be
+    /// - First, select the appropriate bounds check policy for `base`,
+    ///   depending on its address space.
+    ///
+    /// - Next, analyze `index` to see if its value is known at
+    ///   compile time, in which case we can decide statically whether
+    ///   the index is in bounds.
+    ///
+    /// - If the index's value is not known at compile time, emit code to:
+    ///
+    ///     - restrict its value (for [`BoundsCheckPolicy::Restrict`]), or
+    ///
+    ///     - check whether it's in bounds (for
+    ///       [`BoundsCheckPolicy::ReadZeroSkipWrite`]).
+    ///
+    /// Return a [`BoundsCheckResult`] indicating how the index should be
     /// consumed. See that type's documentation for details.
     pub(super) fn write_bounds_check(
         &mut self,

third_party/rust/naga/src/back/spv/mod.rs

@@ -144,6 +144,7 @@ struct Function {
     signature: Option<Instruction>,
     parameters: Vec<FunctionArgument>,
     variables: crate::FastHashMap<Handle<crate::LocalVariable>, LocalVariable>,
+    internal_variables: Vec<LocalVariable>,
     blocks: Vec<TerminatedBlock>,
     entry_point_context: Option<EntryPointContext>,
 }
@@ -466,38 +467,75 @@ enum CachedConstant {
     ZeroValue(Word),
 }
 
+/// The SPIR-V representation of a [`crate::GlobalVariable`].
+///
+/// In the Vulkan spec 1.3.296, the section [Descriptor Set Interface][dsi] says:
+///
+/// > Variables identified with the `Uniform` storage class are used to access
+/// > transparent buffer backed resources. Such variables *must* be:
+/// >
+/// > -   typed as `OpTypeStruct`, or an array of this type,
+/// >
+/// > -   identified with a `Block` or `BufferBlock` decoration, and
+/// >
+/// > -   laid out explicitly using the `Offset`, `ArrayStride`, and `MatrixStride`
+/// >     decorations as specified in "Offset and Stride Assignment".
+///
+/// This is followed by identical language for the `StorageBuffer`,
+/// except that a `BufferBlock` decoration is not allowed.
+///
+/// When we encounter a global variable in the [`Storage`] or [`Uniform`]
+/// address spaces whose type is not already [`Struct`], this backend implicitly
+/// wraps the global variable in a struct: we generate a SPIR-V global variable
+/// holding an `OpTypeStruct` with a single member, whose type is what the Naga
+/// global's type would suggest, decorated as required above.
+///
+/// The [`helpers::global_needs_wrapper`] function determines whether a given
+/// [`crate::GlobalVariable`] needs to be wrapped.
+///
+/// [dsi]: https://registry.khronos.org/vulkan/specs/1.3-extensions/html/vkspec.html#interfaces-resources-descset
+/// [`Storage`]: crate::AddressSpace::Storage
+/// [`Uniform`]: crate::AddressSpace::Uniform
+/// [`Struct`]: crate::TypeInner::Struct
 #[derive(Clone)]
 struct GlobalVariable {
-    /// ID of the OpVariable that declares the global.
+    /// The SPIR-V id of the `OpVariable` that declares the global.
     ///
-    /// If you need the variable's value, use [`access_id`] instead of this
-    /// field. If we wrapped the Naga IR `GlobalVariable`'s type in a struct to
-    /// comply with Vulkan's requirements, then this points to the `OpVariable`
-    /// with the synthesized struct type, whereas `access_id` points to the
-    /// field of said struct that holds the variable's actual value.
+    /// If this global has been implicitly wrapped in an `OpTypeStruct`, this id
+    /// refers to the wrapper, not the original Naga value it contains. If you
+    /// need the Naga value, use [`access_id`] instead of this field.
+    ///
+    /// If this global is not implicitly wrapped, this is the same as
+    /// [`access_id`].
     ///
     /// This is used to compute the `access_id` pointer in function prologues,
-    /// and used for `ArrayLength` expressions, which do need the struct.
+    /// and used for `ArrayLength` expressions, which need to pass the wrapper
+    /// struct.
     ///
     /// [`access_id`]: GlobalVariable::access_id
     var_id: Word,
 
-    /// For `AddressSpace::Handle` variables, this ID is recorded in the function
-    /// prelude block (and reset before every function) as `OpLoad` of the variable.
-    /// It is then used for all the global ops, such as `OpImageSample`.
+    /// The loaded value of a `AddressSpace::Handle` global variable.
+    ///
+    /// If the current function uses this global variable, this is the id of an
+    /// `OpLoad` instruction in the function's prologue that loads its value.
+    /// (This value is assigned as we write the prologue code of each function.)
+    /// It is then used for all operations on the global, such as `OpImageSample`.
     handle_id: Word,
 
-    /// Actual ID used to access this variable.
-    /// For wrapped buffer variables, this ID is `OpAccessChain` into the
-    /// wrapper. Otherwise, the same as `var_id`.
+    /// The SPIR-V id of a pointer to this variable's Naga IR value.
     ///
-    /// Vulkan requires that globals in the `StorageBuffer` and `Uniform` storage
-    /// classes must be structs with the `Block` decoration, but WGSL and Naga IR
-    /// make no such requirement. So for such variables, we generate a wrapper struct
-    /// type with a single element of the type given by Naga, generate an
-    /// `OpAccessChain` for that member in the function prelude, and use that pointer
-    /// to refer to the global in the function body. This is the id of that access,
-    /// updated for each function in `write_function`.
+    /// If the current function uses this global variable, and it has been
+    /// implicitly wrapped in an `OpTypeStruct`, this is the id of an
+    /// `OpAccessChain` instruction in the function's prologue that refers to
+    /// the wrapped value inside the struct. (This value is assigned as we write
+    /// the prologue code of each function.) If you need the wrapper struct
+    /// itself, use [`var_id`] instead of this field.
+    ///
+    /// If this global is not implicitly wrapped, this is the same as
+    /// [`var_id`].
+    ///
+    /// [`var_id`]: GlobalVariable::var_id
     access_id: Word,
 }
 
@@ -627,12 +665,6 @@ impl BlockContext<'_> {
     }
 }
 
-#[derive(Clone, Copy, Default)]
-struct LoopContext {
-    continuing_id: Option<Word>,
-    break_id: Option<Word>,
-}
-
 pub struct Writer {
     physical_layout: PhysicalLayout,
     logical_layout: LogicalLayout,

third_party/rust/naga/src/back/spv/writer.rs

@@ -3,7 +3,7 @@ use super::{
     helpers::{contains_builtin, global_needs_wrapper, map_storage_class},
     make_local, Block, BlockContext, CachedConstant, CachedExpressions, DebugInfo,
     EntryPointContext, Error, Function, FunctionArgument, GlobalVariable, IdGenerator, Instruction,
-    LocalType, LocalVariable, LogicalLayout, LookupFunctionType, LookupType, LoopContext, Options,
+    LocalType, LocalVariable, LogicalLayout, LookupFunctionType, LookupType, Options,
     PhysicalLayout, PipelineOptions, ResultMember, Writer, WriterFlags, BITS_PER_BYTE,
 };
 use crate::{
@@ -32,6 +32,9 @@ impl Function {
                 for local_var in self.variables.values() {
                     local_var.instruction.to_words(sink);
                 }
+                for internal_var in self.internal_variables.iter() {
+                    internal_var.instruction.to_words(sink);
+                }
             }
             for instruction in block.body.iter() {
                 instruction.to_words(sink);
@@ -135,6 +138,56 @@ impl Writer {
         self.capabilities_used.insert(spirv::Capability::Shader);
     }
 
+    #[allow(clippy::too_many_arguments)]
+    pub(super) fn promote_access_expression_to_variable(
+        &mut self,
+        ir_types: &UniqueArena<crate::Type>,
+        result_type_id: Word,
+        container_id: Word,
+        container_ty: Handle<crate::Type>,
+        index_id: Word,
+        element_ty: Handle<crate::Type>,
+        block: &mut Block,
+    ) -> Result<(Word, LocalVariable), Error> {
+        let pointer_type_id =
+            self.get_pointer_id(ir_types, container_ty, spirv::StorageClass::Function)?;
+
+        let variable = {
+            let id = self.id_gen.next();
+            LocalVariable {
+                id,
+                instruction: Instruction::variable(
+                    pointer_type_id,
+                    id,
+                    spirv::StorageClass::Function,
+                    None,
+                ),
+            }
+        };
+        block
+            .body
+            .push(Instruction::store(variable.id, container_id, None));
+
+        let element_pointer_id = self.id_gen.next();
+        let element_pointer_type_id =
+            self.get_pointer_id(ir_types, element_ty, spirv::StorageClass::Function)?;
+        block.body.push(Instruction::access_chain(
+            element_pointer_type_id,
+            element_pointer_id,
+            variable.id,
+            &[index_id],
+        ));
+        let id = self.id_gen.next();
+        block.body.push(Instruction::load(
+            result_type_id,
+            id,
+            element_pointer_id,
+            None,
+        ));
+
+        Ok((id, variable))
+    }
+
     /// Indicate that the code requires any one of the listed capabilities.
     ///
     /// If nothing in `capabilities` appears in the available capabilities
@@ -703,13 +756,7 @@ impl Writer {
             next_id
         };
 
-        context.write_block(
-            main_id,
-            &ir_function.body,
-            super::block::BlockExit::Return,
-            LoopContext::default(),
-            debug_info.as_ref(),
-        )?;
+        context.write_function_body(main_id, debug_info.as_ref())?;
 
         // Consume the `BlockContext`, ending its borrows and letting the
         // `Writer` steal back its cached expression table and temp_list.

third_party/rust/naga/src/back/wgsl/writer.rs

@@ -1221,31 +1221,31 @@ impl<W: Write> Writer<W> {
 
         match expressions[expr] {
             Expression::Literal(literal) => match literal {
-                crate::Literal::F32(value) => write!(self.out, "{}f", value)?,
-                crate::Literal::U32(value) => write!(self.out, "{}u", value)?,
+                crate::Literal::F32(value) => write!(self.out, "{value}f")?,
+                crate::Literal::U32(value) => write!(self.out, "{value}u")?,
                 crate::Literal::I32(value) => {
                     // `-2147483648i` is not valid WGSL. The most negative `i32`
                     // value can only be expressed in WGSL using AbstractInt and
                     // a unary negation operator.
                     if value == i32::MIN {
-                        write!(self.out, "i32({})", value)?;
+                        write!(self.out, "i32({value})")?;
                     } else {
-                        write!(self.out, "{}i", value)?;
+                        write!(self.out, "{value}i")?;
                     }
                 }
-                crate::Literal::Bool(value) => write!(self.out, "{}", value)?,
-                crate::Literal::F64(value) => write!(self.out, "{:?}lf", value)?,
+                crate::Literal::Bool(value) => write!(self.out, "{value}")?,
+                crate::Literal::F64(value) => write!(self.out, "{value:?}lf")?,
                 crate::Literal::I64(value) => {
                     // `-9223372036854775808li` is not valid WGSL. The most negative `i64`
                     // value can only be expressed in WGSL using AbstractInt and
                     // a unary negation operator.
                     if value == i64::MIN {
-                        write!(self.out, "i64({})", value)?;
+                        write!(self.out, "i64({value})")?;
                     } else {
-                        write!(self.out, "{}li", value)?;
+                        write!(self.out, "{value}li")?;
                     }
                 }
-                crate::Literal::U64(value) => write!(self.out, "{:?}lu", value)?,
+                crate::Literal::U64(value) => write!(self.out, "{value:?}lu")?,
                 crate::Literal::AbstractInt(_) | crate::Literal::AbstractFloat(_) => {
                     return Err(Error::Custom(
                         "Abstract types should not appear in IR presented to backends".into(),

third_party/rust/naga/src/front/glsl/context.rs

@@ -630,7 +630,8 @@ impl<'a> Context<'a> {
                             frontend.errors.push(Error {
                                 kind: ErrorKind::SemanticError(
                                     format!(
-                                        "Cannot apply operation to {left_inner:?} and {right_inner:?}"
+                                        "Cannot apply operation to {:?} and {:?}",
+                                        left_inner, right_inner
                                     )
                                     .into(),
                                 ),
@@ -828,7 +829,8 @@ impl<'a> Context<'a> {
                             frontend.errors.push(Error {
                                 kind: ErrorKind::SemanticError(
                                     format!(
-                                        "Cannot apply operation to {left_inner:?} and {right_inner:?}"
+                                        "Cannot apply operation to {:?} and {:?}",
+                                        left_inner, right_inner
                                     )
                                     .into(),
                                 ),
@@ -908,7 +910,8 @@ impl<'a> Context<'a> {
                             frontend.errors.push(Error {
                                 kind: ErrorKind::SemanticError(
                                     format!(
-                                        "Cannot apply operation to {left_inner:?} and {right_inner:?}"
+                                        "Cannot apply operation to {:?} and {:?}",
+                                        left_inner, right_inner
                                     )
                                     .into(),
                                 ),

third_party/rust/naga/src/front/glsl/functions.rs

@@ -634,7 +634,8 @@ impl Frontend {
                         self.errors.push(Error {
                             kind: ErrorKind::SemanticError(
                                 format!(
-                                    "'{name}': image needs {overload_access:?} access but only {call_access:?} was provided"
+                                    "'{}': image needs {:?} access but only {:?} was provided",
+                                    name, overload_access, call_access
                                 )
                                 .into(),
                             ),

third_party/rust/naga/src/front/glsl/parser/expressions.rs

@@ -38,7 +38,13 @@ impl<'source> ParsingContext<'source> {
             TokenValue::FloatConstant(float) => {
                 if float.width != 32 {
                     frontend.errors.push(Error {
-                        kind: ErrorKind::SemanticError("Unsupported floating-point value (expected single-precision floating-point number)".into()),
+                        kind: ErrorKind::SemanticError(
+                            concat!(
+                                "Unsupported floating-point value ",
+                                "(expected single-precision floating-point number)"
+                            )
+                            .into(),
+                        ),
                         meta: token.meta,
                     });
                 }

third_party/rust/naga/src/front/glsl/variables.rs

@@ -294,14 +294,17 @@ impl Frontend {
                             .any(|i| components[i..].contains(&components[i - 1]));
                         if not_unique {
                             self.errors.push(Error {
-                                kind:
-                                ErrorKind::SemanticError(
-                                format!(
-                                    "swizzle cannot have duplicate components in left-hand-side expression for \"{name:?}\""
-                                )
-                                .into(),
-                            ),
-                                meta ,
+                                kind: ErrorKind::SemanticError(
+                                    format!(
+                                        concat!(
+                                            "swizzle cannot have duplicate components in ",
+                                            "left-hand-side expression for \"{:?}\""
+                                        ),
+                                        name
+                                    )
+                                    .into(),
+                                ),
+                                meta,
                             })
                         }
                     }

third_party/rust/naga/src/front/spv/error.rs

@@ -47,7 +47,13 @@ pub enum Error {
     UnsupportedBinaryOperator(spirv::Word),
     #[error("Naga supports OpTypeRuntimeArray in the StorageBuffer storage class only")]
     UnsupportedRuntimeArrayStorageClass,
-    #[error("unsupported matrix stride {stride} for a {columns}x{rows} matrix with scalar width={width}")]
+    #[error(
+        "unsupported matrix stride {} for a {}x{} matrix with scalar width={}",
+        stride,
+        columns,
+        rows,
+        width
+    )]
     UnsupportedMatrixStride {
         stride: u32,
         columns: u8,
@@ -162,6 +168,6 @@ impl Error {
 
 impl From<atomic_upgrade::Error> for Error {
     fn from(source: atomic_upgrade::Error) -> Self {
-        crate::front::spv::Error::AtomicUpgradeError(source)
+        Error::AtomicUpgradeError(source)
     }
 }

third_party/rust/naga/src/front/wgsl/error.rs

@@ -298,32 +298,42 @@ impl<'a> Error<'a> {
         match *self {
             Error::Unexpected(unexpected_span, expected) => {
                 let expected_str = match expected {
-                    ExpectedToken::Token(token) => {
-                        match token {
-                            Token::Separator(c) => format!("'{c}'"),
-                            Token::Paren(c) => format!("'{c}'"),
-                            Token::Attribute => "@".to_string(),
-                            Token::Number(_) => "number".to_string(),
-                            Token::Word(s) => s.to_string(),
-                            Token::Operation(c) => format!("operation ('{c}')"),
-                            Token::LogicalOperation(c) => format!("logical operation ('{c}')"),
-                            Token::ShiftOperation(c) => format!("bitshift ('{c}{c}')"),
-                            Token::AssignmentOperation(c) if c=='<' || c=='>' => format!("bitshift ('{c}{c}=')"),
-                            Token::AssignmentOperation(c) => format!("operation ('{c}=')"),
-                            Token::IncrementOperation => "increment operation".to_string(),
-                            Token::DecrementOperation => "decrement operation".to_string(),
-                            Token::Arrow => "->".to_string(),
-                            Token::Unknown(c) => format!("unknown ('{c}')"),
-                            Token::Trivia => "trivia".to_string(),
-                            Token::End => "end".to_string(),
+                    ExpectedToken::Token(token) => match token {
+                        Token::Separator(c) => format!("'{c}'"),
+                        Token::Paren(c) => format!("'{c}'"),
+                        Token::Attribute => "@".to_string(),
+                        Token::Number(_) => "number".to_string(),
+                        Token::Word(s) => s.to_string(),
+                        Token::Operation(c) => format!("operation ('{c}')"),
+                        Token::LogicalOperation(c) => format!("logical operation ('{c}')"),
+                        Token::ShiftOperation(c) => format!("bitshift ('{c}{c}')"),
+                        Token::AssignmentOperation(c) if c == '<' || c == '>' => {
+                            format!("bitshift ('{c}{c}=')")
                         }
-                    }
+                        Token::AssignmentOperation(c) => format!("operation ('{c}=')"),
+                        Token::IncrementOperation => "increment operation".to_string(),
+                        Token::DecrementOperation => "decrement operation".to_string(),
+                        Token::Arrow => "->".to_string(),
+                        Token::Unknown(c) => format!("unknown ('{c}')"),
+                        Token::Trivia => "trivia".to_string(),
+                        Token::End => "end".to_string(),
+                    },
                     ExpectedToken::Identifier => "identifier".to_string(),
                     ExpectedToken::PrimaryExpression => "expression".to_string(),
                     ExpectedToken::Assignment => "assignment or increment/decrement".to_string(),
-                    ExpectedToken::SwitchItem => "switch item ('case' or 'default') or a closing curly bracket to signify the end of the switch statement ('}')".to_string(),
-                    ExpectedToken::WorkgroupSizeSeparator => "workgroup size separator (',') or a closing parenthesis".to_string(),
-                    ExpectedToken::GlobalItem => "global item ('struct', 'const', 'var', 'alias', ';', 'fn') or the end of the file".to_string(),
+                    ExpectedToken::SwitchItem => concat!(
+                        "switch item ('case' or 'default') or a closing curly bracket ",
+                        "to signify the end of the switch statement ('}')"
+                    )
+                    .to_string(),
+                    ExpectedToken::WorkgroupSizeSeparator => {
+                        "workgroup size separator (',') or a closing parenthesis".to_string()
+                    }
+                    ExpectedToken::GlobalItem => concat!(
+                        "global item ('struct', 'const', 'var', 'alias', ';', 'fn') ",
+                        "or the end of the file"
+                    )
+                    .to_string(),
                     ExpectedToken::Type => "type".to_string(),
                     ExpectedToken::Variable => "variable access".to_string(),
                     ExpectedToken::Function => "function name".to_string(),
@@ -384,9 +394,11 @@ impl<'a> Error<'a> {
                 notes: vec![],
             },
             Error::BadIncrDecrReferenceType(span) => ParseError {
-                message:
-                    "increment/decrement operation requires reference type to be one of i32 or u32"
-                        .to_string(),
+                message: concat!(
+                    "increment/decrement operation requires ",
+                    "reference type to be one of i32 or u32"
+                )
+                .to_string(),
                 labels: vec![(span, "must be a reference type of i32 or u32".into())],
                 notes: vec![],
             },
@@ -527,25 +539,24 @@ impl<'a> Error<'a> {
                 labels: vec![(span, "type can't be inferred".into())],
                 notes: vec![],
             },
-            Error::InitializationTypeMismatch { name, ref expected, ref got } => {
-                ParseError {
-                    message: format!(
-                        "the type of `{}` is expected to be `{}`, but got `{}`",
-                        &source[name], expected, got,
-                    ),
-                    labels: vec![(
-                        name,
-                        format!("definition of `{}`", &source[name]).into(),
-                    )],
-                    notes: vec![],
-                }
-            }
+            Error::InitializationTypeMismatch {
+                name,
+                ref expected,
+                ref got,
+            } => ParseError {
+                message: format!(
+                    "the type of `{}` is expected to be `{}`, but got `{}`",
+                    &source[name], expected, got,
+                ),
+                labels: vec![(name, format!("definition of `{}`", &source[name]).into())],
+                notes: vec![],
+            },
             Error::DeclMissingTypeAndInit(name_span) => ParseError {
-                message: format!("declaration of `{}` needs a type specifier or initializer", &source[name_span]),
-                labels: vec![(
-                    name_span,
-                    "needs a type specifier or initializer".into(),
-                )],
+                message: format!(
+                    "declaration of `{}` needs a type specifier or initializer",
+                    &source[name_span]
+                ),
+                labels: vec![(name_span, "needs a type specifier or initializer".into())],
                 notes: vec![],
             },
             Error::MissingAttribute(name, name_span) => ParseError {
@@ -725,7 +736,11 @@ impl<'a> Error<'a> {
                 notes: vec![message.into()],
             },
             Error::ExpectedConstExprConcreteIntegerScalar(span) => ParseError {
-                message: "must be a const-expression that resolves to a concrete integer scalar (u32 or i32)".to_string(),
+                message: concat!(
+                    "must be a const-expression that ",
+                    "resolves to a concrete integer scalar (u32 or i32)"
+                )
+                .to_string(),
                 labels: vec![(span, "must resolve to u32 or i32".into())],
                 notes: vec![],
             },
@@ -754,9 +769,17 @@ impl<'a> Error<'a> {
             },
             Error::AutoConversion(ref error) => {
                 // destructuring ensures all fields are handled
-                let AutoConversionError { dest_span, ref dest_type, source_span, ref source_type } = **error;
+                let AutoConversionError {
+                    dest_span,
+                    ref dest_type,
+                    source_span,
+                    ref source_type,
+                } = **error;
                 ParseError {
-                    message: format!("automatic conversions cannot convert `{source_type}` to `{dest_type}`"),
+                    message: format!(
+                        "automatic conversions cannot convert `{}` to `{}`",
+                        source_type, dest_type
+                    ),
                     labels: vec![
                         (
                             dest_span,
@@ -765,72 +788,77 @@ impl<'a> Error<'a> {
                         (
                             source_span,
                             format!("this expression has type {source_type}").into(),
-                        )
+                        ),
                     ],
                     notes: vec![],
                 }
-            },
+            }
             Error::AutoConversionLeafScalar(ref error) => {
-                let AutoConversionLeafScalarError { dest_span, ref dest_scalar, source_span, ref source_type } = **error;
+                let AutoConversionLeafScalarError {
+                    dest_span,
+                    ref dest_scalar,
+                    source_span,
+                    ref source_type,
+                } = **error;
                 ParseError {
-                    message: format!("automatic conversions cannot convert elements of `{source_type}` to `{dest_scalar}`"),
+                    message: format!(
+                        "automatic conversions cannot convert elements of `{}` to `{}`",
+                        source_type, dest_scalar
+                    ),
                     labels: vec![
                         (
                             dest_span,
-                            format!("a value with elements of type {dest_scalar} is required here").into(),
+                            format!(
+                                "a value with elements of type {} is required here",
+                                dest_scalar
+                            )
+                            .into(),
                         ),
                         (
                             source_span,
                             format!("this expression has type {source_type}").into(),
-                        )
+                        ),
                     ],
                     notes: vec![],
                 }
-            },
+            }
             Error::ConcretizationFailed(ref error) => {
-                let ConcretizationFailedError { expr_span, ref expr_type, ref scalar, ref inner } = **error;
+                let ConcretizationFailedError {
+                    expr_span,
+                    ref expr_type,
+                    ref scalar,
+                    ref inner,
+                } = **error;
                 ParseError {
-                    message: format!("failed to convert expression to a concrete type: {}", inner),
-                    labels: vec![
-                        (
-                            expr_span,
-                            format!("this expression has type {}", expr_type).into(),
-                        )
-                    ],
-                    notes: vec![
-                        format!("the expression should have been converted to have {} scalar type", scalar),
-                    ]
+                    message: format!("failed to convert expression to a concrete type: {inner}"),
+                    labels: vec![(
+                        expr_span,
+                        format!("this expression has type {expr_type}").into(),
+                    )],
+                    notes: vec![format!(
+                        "the expression should have been converted to have {} scalar type",
+                        scalar
+                    )],
                 }
-            },
+            }
             Error::ExceededLimitForNestedBraces { span, limit } => ParseError {
                 message: "brace nesting limit reached".into(),
                 labels: vec![(span, "limit reached at this brace".into())],
-                notes: vec![
-                    format!("nesting limit is currently set to {limit}"),
-                ],
+                notes: vec![format!("nesting limit is currently set to {limit}")],
             },
             Error::PipelineConstantIDValue(span) => ParseError {
                 message: "pipeline constant ID must be between 0 and 65535 inclusive".to_string(),
-                labels: vec![(
-                    span,
-                    "must be between 0 and 65535 inclusive".into(),
-                )],
+                labels: vec![(span, "must be between 0 and 65535 inclusive".into())],
                 notes: vec![],
             },
             Error::NotBool(span) => ParseError {
                 message: "must be a const-expression that resolves to a bool".to_string(),
-                labels: vec![(
-                    span,
-                    "must resolve to bool".into(),
-                )],
+                labels: vec![(span, "must resolve to bool".into())],
                 notes: vec![],
             },
             Error::ConstAssertFailed(span) => ParseError {
                 message: "const_assert failure".to_string(),
-                labels: vec![(
-                    span,
-                    "evaluates to false".into(),
-                )],
+                labels: vec![(span, "evaluates to false".into())],
                 notes: vec![],
             },
         }

third_party/rust/naga/src/front/wgsl/lower/mod.rs

@@ -1778,12 +1778,14 @@ impl<'source, 'temp> Lowerer<'source, 'temp> {
 
                 return Ok(());
             }
-            ast::StatementKind::Ignore(expr) => {
+            ast::StatementKind::Phony(expr) => {
                 let mut emitter = Emitter::default();
                 emitter.start(&ctx.function.expressions);
 
-                let _ = self.expression(expr, &mut ctx.as_expression(block, &mut emitter))?;
+                let value = self.expression(expr, &mut ctx.as_expression(block, &mut emitter))?;
                 block.extend(emitter.finish(&ctx.function.expressions));
+                ctx.named_expressions
+                    .insert(value, ("phony".to_string(), stmt.span));
                 return Ok(());
             }
         };

third_party/rust/naga/src/front/wgsl/parse/ast.rs

@@ -284,7 +284,7 @@ pub enum StatementKind<'a> {
     },
     Increment(Handle<Expression<'a>>),
     Decrement(Handle<Expression<'a>>),
-    Ignore(Handle<Expression<'a>>),
+    Phony(Handle<Expression<'a>>),
     ConstAssert(Handle<Expression<'a>>),
 }
 

third_party/rust/naga/src/front/wgsl/parse/mod.rs

@@ -1696,7 +1696,7 @@ impl Parser {
                         let expr = self.general_expression(lexer, ctx)?;
                         lexer.expect(Token::Separator(';'))?;
 
-                        ast::StatementKind::Ignore(expr)
+                        ast::StatementKind::Phony(expr)
                     }
                     "let" => {
                         let _ = lexer.next();

third_party/rust/naga/src/lib.rs

@@ -1402,21 +1402,20 @@ pub enum Expression {
     /// ## Dynamic indexing restrictions
     ///
     /// To accommodate restrictions in some of the shader languages that Naga
-    /// targets, it is not permitted to subscript a matrix or array with a
-    /// dynamically computed index unless that matrix or array appears behind a
-    /// pointer. In other words, if the inner type of `base` is [`Array`] or
-    /// [`Matrix`], then `index` must be a constant. But if the type of `base`
-    /// is a [`Pointer`] to an array or matrix or a [`ValuePointer`] with a
-    /// `size`, then the index may be any expression of integer type.
+    /// targets, it is not permitted to subscript a matrix with a dynamically
+    /// computed index unless that matrix appears behind a pointer. In other
+    /// words, if the inner type of `base` is [`Matrix`], then `index` must be a
+    /// constant. But if the type of `base` is a [`Pointer`] to an matrix, then
+    /// the index may be any expression of integer type.
     ///
     /// You can use the [`Expression::is_dynamic_index`] method to determine
-    /// whether a given index expression requires matrix or array base operands
-    /// to be behind a pointer.
+    /// whether a given index expression requires matrix base operands to be
+    /// behind a pointer.
     ///
     /// (It would be simpler to always require the use of `AccessIndex` when
-    /// subscripting arrays and matrices that are not behind pointers, but to
-    /// accommodate existing front ends, Naga also permits `Access`, with a
-    /// restricted `index`.)
+    /// subscripting matrices that are not behind pointers, but to accommodate
+    /// existing front ends, Naga also permits `Access`, with a restricted
+    /// `index`.)
     ///
     /// [`Vector`]: TypeInner::Vector
     /// [`Matrix`]: TypeInner::Matrix

third_party/rust/naga/src/proc/mod.rs

@@ -521,12 +521,12 @@ impl crate::Expression {
         }
     }
 
-    /// Return true if this expression is a dynamic array index, for [`Access`].
+    /// Return true if this expression is a dynamic array/vector/matrix index,
+    /// for [`Access`].
     ///
     /// This method returns true if this expression is a dynamically computed
-    /// index, and as such can only be used to index matrices and arrays when
-    /// they appear behind a pointer. See the documentation for [`Access`] for
-    /// details.
+    /// index, and as such can only be used to index matrices when they appear
+    /// behind a pointer. See the documentation for [`Access`] for details.
     ///
     /// Note, this does not check the _type_ of the given expression. It's up to
     /// the caller to establish that the `Access` expression is well-typed

third_party/rust/naga/src/proc/typifier.rs

@@ -92,6 +92,13 @@ pub enum TypeResolution {
     /// available in the associated arena. However, the `TypeInner` itself may
     /// contain `Handle<Type>` values referring to types from the arena.
     ///
+    /// The inner type must only be one of the following variants:
+    /// - TypeInner::Pointer
+    /// - TypeInner::ValuePointer
+    /// - TypeInner::Matrix (generated by matrix multiplication)
+    /// - TypeInner::Vector
+    /// - TypeInner::Scalar
+    ///
     /// [`TypeInner`]: crate::TypeInner
     Value(crate::TypeInner),
 }
@@ -631,41 +638,37 @@ impl<'a> ResolveContext<'a> {
                 use crate::MathFunction as Mf;
                 let res_arg = past(arg)?;
                 match fun {
-                    // comparison
-                    Mf::Abs |
-                    Mf::Min |
-                    Mf::Max |
-                    Mf::Clamp |
-                    Mf::Saturate |
-                    // trigonometry
-                    Mf::Cos |
-                    Mf::Cosh |
-                    Mf::Sin |
-                    Mf::Sinh |
-                    Mf::Tan |
-                    Mf::Tanh |
-                    Mf::Acos |
-                    Mf::Asin |
-                    Mf::Atan |
-                    Mf::Atan2 |
-                    Mf::Asinh |
-                    Mf::Acosh |
-                    Mf::Atanh |
-                    Mf::Radians |
-                    Mf::Degrees |
-                    // decomposition
-                    Mf::Ceil |
-                    Mf::Floor |
-                    Mf::Round |
-                    Mf::Fract |
-                    Mf::Trunc |
-                    Mf::Ldexp |
-                    // exponent
-                    Mf::Exp |
-                    Mf::Exp2 |
-                    Mf::Log |
-                    Mf::Log2 |
-                    Mf::Pow => res_arg.clone(),
+                    Mf::Abs
+                    | Mf::Min
+                    | Mf::Max
+                    | Mf::Clamp
+                    | Mf::Saturate
+                    | Mf::Cos
+                    | Mf::Cosh
+                    | Mf::Sin
+                    | Mf::Sinh
+                    | Mf::Tan
+                    | Mf::Tanh
+                    | Mf::Acos
+                    | Mf::Asin
+                    | Mf::Atan
+                    | Mf::Atan2
+                    | Mf::Asinh
+                    | Mf::Acosh
+                    | Mf::Atanh
+                    | Mf::Radians
+                    | Mf::Degrees
+                    | Mf::Ceil
+                    | Mf::Floor
+                    | Mf::Round
+                    | Mf::Fract
+                    | Mf::Trunc
+                    | Mf::Ldexp
+                    | Mf::Exp
+                    | Mf::Exp2
+                    | Mf::Log
+                    | Mf::Log2
+                    | Mf::Pow => res_arg.clone(),
                     Mf::Modf | Mf::Frexp => {
                         let (size, width) = match res_arg.inner_with(types) {
                             &Ti::Scalar(crate::Scalar {
@@ -673,77 +676,81 @@ impl<'a> ResolveContext<'a> {
                                 width,
                             }) => (None, width),
                             &Ti::Vector {
-                                scalar: crate::Scalar {
-                                    kind: crate::ScalarKind::Float,
-                                    width,
-                                },
+                                scalar:
+                                    crate::Scalar {
+                                        kind: crate::ScalarKind::Float,
+                                        width,
+                                    },
                                 size,
                             } => (Some(size), width),
-                            ref other =>
-                                return Err(ResolveError::IncompatibleOperands(format!("{fun:?}({other:?}, _)")))
+                            ref other => {
+                                return Err(ResolveError::IncompatibleOperands(format!(
+                                    "{fun:?}({other:?}, _)"
+                                )))
+                            }
                         };
                         let result = self
-                        .special_types
-                        .predeclared_types
-                        .get(&if fun == Mf::Modf {
-                            crate::PredeclaredType::ModfResult { size, width }
-                    } else {
-                            crate::PredeclaredType::FrexpResult { size, width }
-                    })
-                        .ok_or(ResolveError::MissingSpecialType)?;
+                            .special_types
+                            .predeclared_types
+                            .get(&if fun == Mf::Modf {
+                                crate::PredeclaredType::ModfResult { size, width }
+                            } else {
+                                crate::PredeclaredType::FrexpResult { size, width }
+                            })
+                            .ok_or(ResolveError::MissingSpecialType)?;
                         TypeResolution::Handle(*result)
-                    },
-                    // geometry
+                    }
                     Mf::Dot => match *res_arg.inner_with(types) {
-                        Ti::Vector {
-                            size: _,
-                            scalar,
-                        } => TypeResolution::Value(Ti::Scalar(scalar)),
-                        ref other =>
-                            return Err(ResolveError::IncompatibleOperands(
-                                format!("{fun:?}({other:?}, _)")
-                            )),
+                        Ti::Vector { size: _, scalar } => TypeResolution::Value(Ti::Scalar(scalar)),
+                        ref other => {
+                            return Err(ResolveError::IncompatibleOperands(format!(
+                                "{fun:?}({other:?}, _)"
+                            )))
+                        }
                     },
                     Mf::Outer => {
-                        let arg1 = arg1.ok_or_else(|| ResolveError::IncompatibleOperands(
-                            format!("{fun:?}(_, None)")
-                        ))?;
+                        let arg1 = arg1.ok_or_else(|| {
+                            ResolveError::IncompatibleOperands(format!("{fun:?}(_, None)"))
+                        })?;
                         match (res_arg.inner_with(types), past(arg1)?.inner_with(types)) {
                             (
-                                &Ti::Vector { size: columns, scalar },
-                                &Ti::Vector{ size: rows, .. }
+                                &Ti::Vector {
+                                    size: columns,
+                                    scalar,
+                                },
+                                &Ti::Vector { size: rows, .. },
                             ) => TypeResolution::Value(Ti::Matrix {
                                 columns,
                                 rows,
                                 scalar,
                             }),
-                            (left, right) =>
-                                return Err(ResolveError::IncompatibleOperands(
-                                    format!("{fun:?}({left:?}, {right:?})")
-                                )),
+                            (left, right) => {
+                                return Err(ResolveError::IncompatibleOperands(format!(
+                                    "{fun:?}({left:?}, {right:?})"
+                                )))
+                            }
+                        }
+                    }
+                    Mf::Cross => res_arg.clone(),
+                    Mf::Distance | Mf::Length => match *res_arg.inner_with(types) {
+                        Ti::Scalar(scalar) | Ti::Vector { scalar, size: _ } => {
+                            TypeResolution::Value(Ti::Scalar(scalar))
+                        }
+                        ref other => {
+                            return Err(ResolveError::IncompatibleOperands(format!(
+                                "{fun:?}({other:?})"
+                            )))
                         }
                     },
-                    Mf::Cross => res_arg.clone(),
-                    Mf::Distance |
-                    Mf::Length => match *res_arg.inner_with(types) {
-                        Ti::Scalar(scalar) |
-                        Ti::Vector {scalar,size:_} => TypeResolution::Value(Ti::Scalar(scalar)),
-                        ref other => return Err(ResolveError::IncompatibleOperands(
-                                format!("{fun:?}({other:?})")
-                            )),
-                    },
-                    Mf::Normalize |
-                    Mf::FaceForward |
-                    Mf::Reflect |
-                    Mf::Refract => res_arg.clone(),
+                    Mf::Normalize | Mf::FaceForward | Mf::Reflect | Mf::Refract => res_arg.clone(),
                     // computational
-                    Mf::Sign |
-                    Mf::Fma |
-                    Mf::Mix |
-                    Mf::Step |
-                    Mf::SmoothStep |
-                    Mf::Sqrt |
-                    Mf::InverseSqrt => res_arg.clone(),
+                    Mf::Sign
+                    | Mf::Fma
+                    | Mf::Mix
+                    | Mf::Step
+                    | Mf::SmoothStep
+                    | Mf::Sqrt
+                    | Mf::InverseSqrt => res_arg.clone(),
                     Mf::Transpose => match *res_arg.inner_with(types) {
                         Ti::Matrix {
                             columns,
@@ -754,9 +761,11 @@ impl<'a> ResolveContext<'a> {
                             rows: columns,
                             scalar,
                         }),
-                        ref other => return Err(ResolveError::IncompatibleOperands(
-                            format!("{fun:?}({other:?})")
-                        )),
+                        ref other => {
+                            return Err(ResolveError::IncompatibleOperands(format!(
+                                "{fun:?}({other:?})"
+                            )))
+                        }
                     },
                     Mf::Inverse => match *res_arg.inner_with(types) {
                         Ti::Matrix {
@@ -768,70 +777,75 @@ impl<'a> ResolveContext<'a> {
                             rows,
                             scalar,
                         }),
-                        ref other => return Err(ResolveError::IncompatibleOperands(
-                            format!("{fun:?}({other:?})")
-                        )),
+                        ref other => {
+                            return Err(ResolveError::IncompatibleOperands(format!(
+                                "{fun:?}({other:?})"
+                            )))
+                        }
                     },
                     Mf::Determinant => match *res_arg.inner_with(types) {
-                        Ti::Matrix {
-                            scalar,
-                            ..
-                        } => TypeResolution::Value(Ti::Scalar(scalar)),
-                        ref other => return Err(ResolveError::IncompatibleOperands(
-                            format!("{fun:?}({other:?})")
-                        )),
+                        Ti::Matrix { scalar, .. } => TypeResolution::Value(Ti::Scalar(scalar)),
+                        ref other => {
+                            return Err(ResolveError::IncompatibleOperands(format!(
+                                "{fun:?}({other:?})"
+                            )))
+                        }
                     },
                     // bits
-                    Mf::CountTrailingZeros |
-                    Mf::CountLeadingZeros |
-                    Mf::CountOneBits |
-                    Mf::ReverseBits |
-                    Mf::ExtractBits |
-                    Mf::InsertBits |
-                    Mf::FirstTrailingBit |
-                    Mf::FirstLeadingBit => match *res_arg.inner_with(types)  {
-                        Ti::Scalar(scalar @ crate::Scalar {
-                            kind: crate::ScalarKind::Sint | crate::ScalarKind::Uint,
-                            ..
-                        }) => TypeResolution::Value(Ti::Scalar(scalar)),
-                        Ti::Vector {
-                            size,
-                            scalar: scalar @ crate::Scalar {
+                    Mf::CountTrailingZeros
+                    | Mf::CountLeadingZeros
+                    | Mf::CountOneBits
+                    | Mf::ReverseBits
+                    | Mf::ExtractBits
+                    | Mf::InsertBits
+                    | Mf::FirstTrailingBit
+                    | Mf::FirstLeadingBit => match *res_arg.inner_with(types) {
+                        Ti::Scalar(
+                            scalar @ crate::Scalar {
                                 kind: crate::ScalarKind::Sint | crate::ScalarKind::Uint,
                                 ..
-                            }
+                            },
+                        ) => TypeResolution::Value(Ti::Scalar(scalar)),
+                        Ti::Vector {
+                            size,
+                            scalar:
+                                scalar @ crate::Scalar {
+                                    kind: crate::ScalarKind::Sint | crate::ScalarKind::Uint,
+                                    ..
+                                },
                         } => TypeResolution::Value(Ti::Vector { size, scalar }),
-                        ref other => return Err(ResolveError::IncompatibleOperands(
-                                format!("{fun:?}({other:?})")
-                            )),
+                        ref other => {
+                            return Err(ResolveError::IncompatibleOperands(format!(
+                                "{fun:?}({other:?})"
+                            )))
+                        }
                     },
                     // data packing
-                    Mf::Pack4x8snorm |
-                    Mf::Pack4x8unorm |
-                    Mf::Pack2x16snorm |
-                    Mf::Pack2x16unorm |
-                    Mf::Pack2x16float |
-                    Mf::Pack4xI8 |
-                    Mf::Pack4xU8 => TypeResolution::Value(Ti::Scalar(crate::Scalar::U32)),
+                    Mf::Pack4x8snorm
+                    | Mf::Pack4x8unorm
+                    | Mf::Pack2x16snorm
+                    | Mf::Pack2x16unorm
+                    | Mf::Pack2x16float
+                    | Mf::Pack4xI8
+                    | Mf::Pack4xU8 => TypeResolution::Value(Ti::Scalar(crate::Scalar::U32)),
                     // data unpacking
-                    Mf::Unpack4x8snorm |
-                    Mf::Unpack4x8unorm => TypeResolution::Value(Ti::Vector {
+                    Mf::Unpack4x8snorm | Mf::Unpack4x8unorm => TypeResolution::Value(Ti::Vector {
                         size: crate::VectorSize::Quad,
-                        scalar: crate::Scalar::F32
-                    }),
-                    Mf::Unpack2x16snorm |
-                    Mf::Unpack2x16unorm |
-                    Mf::Unpack2x16float => TypeResolution::Value(Ti::Vector {
-                        size: crate::VectorSize::Bi,
-                        scalar: crate::Scalar::F32
+                        scalar: crate::Scalar::F32,
                     }),
+                    Mf::Unpack2x16snorm | Mf::Unpack2x16unorm | Mf::Unpack2x16float => {
+                        TypeResolution::Value(Ti::Vector {
+                            size: crate::VectorSize::Bi,
+                            scalar: crate::Scalar::F32,
+                        })
+                    }
                     Mf::Unpack4xI8 => TypeResolution::Value(Ti::Vector {
                         size: crate::VectorSize::Quad,
-                        scalar: crate::Scalar::I32
+                        scalar: crate::Scalar::I32,
                     }),
                     Mf::Unpack4xU8 => TypeResolution::Value(Ti::Vector {
                         size: crate::VectorSize::Quad,
-                        scalar: crate::Scalar::U32
+                        scalar: crate::Scalar::U32,
                     }),
                 }
             }

third_party/rust/naga/src/span.rs

@@ -11,6 +11,7 @@ pub struct Span {
 
 impl Span {
     pub const UNDEFINED: Self = Self { start: 0, end: 0 };
+
     /// Creates a new `Span` from a range of byte indices
     ///
     /// Note: end is exclusive, it doesn't belong to the `Span`

third_party/rust/naga/src/valid/analyzer.rs

@@ -589,23 +589,16 @@ impl FunctionInfo {
                     requirements: UniformityRequirements::empty(),
                 }
             }
-            // depends on the builtin or interpolation
+            // depends on the builtin
             E::FunctionArgument(index) => {
                 let arg = &resolve_context.arguments[index as usize];
                 let uniform = match arg.binding {
                     Some(crate::Binding::BuiltIn(
-                        // per-polygon built-ins are uniform
-                        crate::BuiltIn::FrontFacing
                         // per-work-group built-ins are uniform
-                        | crate::BuiltIn::WorkGroupId
+                        crate::BuiltIn::WorkGroupId
                         | crate::BuiltIn::WorkGroupSize
-                        | crate::BuiltIn::NumWorkGroups)
-                    ) => true,
-                    // only flat inputs are uniform
-                    Some(crate::Binding::Location {
-                        interpolation: Some(crate::Interpolation::Flat),
-                        ..
-                    }) => true,
+                        | crate::BuiltIn::NumWorkGroups,
+                    )) => true,
                     _ => false,
                 };
                 Uniformity {

third_party/rust/naga/src/valid/expression.rs

@@ -240,9 +240,10 @@ impl super::Validator {
                 let base_type = &resolver[base];
                 // See the documentation for `Expression::Access`.
                 let dynamic_indexing_restricted = match *base_type {
-                    Ti::Vector { .. } => false,
-                    Ti::Matrix { .. } | Ti::Array { .. } => true,
-                    Ti::Pointer { .. }
+                    Ti::Matrix { .. } => true,
+                    Ti::Vector { .. }
+                    | Ti::Array { .. }
+                    | Ti::Pointer { .. }
                     | Ti::ValuePointer { size: Some(_), .. }
                     | Ti::BindingArray { .. } => false,
                     ref other => {

third_party/rust/naga/src/valid/type.rs

@@ -664,9 +664,6 @@ impl super::Validator {
                 )
             }
             Ti::BindingArray { base, size } => {
-                if base >= handle {
-                    return Err(TypeError::InvalidArrayBaseType(base));
-                }
                 let type_info_mask = match size {
                     crate::ArraySize::Constant(_) => TypeFlags::SIZED | TypeFlags::HOST_SHAREABLE,
                     crate::ArraySize::Dynamic => {
@@ -680,7 +677,6 @@ impl super::Validator {
                     // Currently Naga only supports binding arrays of structs for non-handle types.
                     match gctx.types[base].inner {
                         crate::TypeInner::Struct { .. } => {}
-                        crate::TypeInner::Array { .. } => {}
                         _ => return Err(TypeError::BindingArrayBaseTypeNotStruct(base)),
                     };
                 }

third_party/rust/thiserror-impl/.cargo-checksum.json

@@ -1 +1 @@
-{"files":{"Cargo.toml":"8f4acbe57979cf65fdca303b08b8ba7bda1d71b73dbdf699b5b765842103fd6c","LICENSE-APACHE":"62c7a1e35f56406896d7aa7ca52d0cc0d272ac022b5d2796e7d6905db8a3636a","LICENSE-MIT":"23f18e03dc49df91622fe2a76176497404e46ced8a715d9d2b67a7446571cca3","src/ast.rs":"9b6cd6b1553483c99cd7e36aa422d37f4353c99b15da55534d28822f7fa7fd08","src/attr.rs":"d662f37c1a892aac6ab652ba6111335121748df728e49399aff736b654d5dd1c","src/expand.rs":"2736a714372a4b81ac5438783dd2c0f656d624bec5cc4089af96ceecaeee011e","src/fmt.rs":"5d1cefc012403c2d4ff7ab2513c0ec559166df4271d5983a6463939b5ec8c3e1","src/generics.rs":"2076cde22271be355a8131a77add4b93f83ab0af4317cd2df5471fffa4f95c66","src/lib.rs":"5eea86c771e643328ad9bc3b881cce4bf9d50adae1b33e0d07645bdd9044003d","src/prop.rs":"5ba613e38430831259f20b258f33d57dcb783fbaeeb49e5faffa7b2a7be99e67","src/span.rs":"430460a4fa0d1fa9c627c1ddd575d2b101778fea84217591e1a93a5f6a2a0132","src/valid.rs":"ac95253944fd360d3578d0643a7baabb2cfa6bf9fbced7a6ce1f7b0529a3bb98"},"package":"a4558b58466b9ad7ca0f102865eccc95938dca1a74a856f2b57b6629050da261"}
+{"files":{"Cargo.toml":"8abf73f36897db35f75f1efed6733948434c7bbc78e002b2a824f46a289c5b65","LICENSE-APACHE":"62c7a1e35f56406896d7aa7ca52d0cc0d272ac022b5d2796e7d6905db8a3636a","LICENSE-MIT":"23f18e03dc49df91622fe2a76176497404e46ced8a715d9d2b67a7446571cca3","src/ast.rs":"9b6cd6b1553483c99cd7e36aa422d37f4353c99b15da55534d28822f7fa7fd08","src/attr.rs":"d662f37c1a892aac6ab652ba6111335121748df728e49399aff736b654d5dd1c","src/expand.rs":"50c30146e65a28ac4f6768e5e9d173bde0162b7ad7c5adc39e4eab6e69650371","src/fmt.rs":"5d1cefc012403c2d4ff7ab2513c0ec559166df4271d5983a6463939b5ec8c3e1","src/generics.rs":"2076cde22271be355a8131a77add4b93f83ab0af4317cd2df5471fffa4f95c66","src/lib.rs":"5eea86c771e643328ad9bc3b881cce4bf9d50adae1b33e0d07645bdd9044003d","src/prop.rs":"5ba613e38430831259f20b258f33d57dcb783fbaeeb49e5faffa7b2a7be99e67","src/span.rs":"430460a4fa0d1fa9c627c1ddd575d2b101778fea84217591e1a93a5f6a2a0132","src/valid.rs":"ac95253944fd360d3578d0643a7baabb2cfa6bf9fbced7a6ce1f7b0529a3bb98"},"package":"08904e7672f5eb876eaaf87e0ce17857500934f4981c4a0ab2b4aa98baac7fc3"}

third_party/rust/thiserror-impl/Cargo.toml

@@ -13,7 +13,7 @@
 edition = "2021"
 rust-version = "1.56"
 name = "thiserror-impl"
-version = "1.0.63"
+version = "1.0.64"
 authors = ["David Tolnay <dtolnay@gmail.com>"]
 build = false
 autobins = false

third_party/rust/thiserror-impl/src/expand.rs

@@ -36,6 +36,7 @@ fn fallback(input: &DeriveInput, error: syn::Error) -> TokenStream {
         #error
 
         #[allow(unused_qualifications)]
+        #[automatically_derived]
         impl #impl_generics std::error::Error for #ty #ty_generics #where_clause
         where
             // Work around trivial bounds being unstable.
@@ -44,6 +45,7 @@ fn fallback(input: &DeriveInput, error: syn::Error) -> TokenStream {
         {}
 
         #[allow(unused_qualifications)]
+        #[automatically_derived]
         impl #impl_generics ::core::fmt::Display for #ty #ty_generics #where_clause {
             fn fmt(&self, __formatter: &mut ::core::fmt::Formatter) -> ::core::fmt::Result {
                 ::core::unreachable!()
@@ -178,6 +180,7 @@ fn impl_struct(input: Struct) -> TokenStream {
         let display_where_clause = display_inferred_bounds.augment_where_clause(input.generics);
         quote! {
             #[allow(unused_qualifications)]
+            #[automatically_derived]
             impl #impl_generics ::core::fmt::Display for #ty #ty_generics #display_where_clause {
                 #[allow(clippy::used_underscore_binding)]
                 fn fmt(&self, __formatter: &mut ::core::fmt::Formatter) -> ::core::fmt::Result {
@@ -193,6 +196,7 @@ fn impl_struct(input: Struct) -> TokenStream {
         let body = from_initializer(from_field, backtrace_field);
         quote! {
             #[allow(unused_qualifications)]
+            #[automatically_derived]
             impl #impl_generics ::core::convert::From<#from> for #ty #ty_generics #where_clause {
                 #[allow(deprecated)]
                 fn from(source: #from) -> Self {
@@ -211,6 +215,7 @@ fn impl_struct(input: Struct) -> TokenStream {
 
     quote! {
         #[allow(unused_qualifications)]
+        #[automatically_derived]
         impl #impl_generics std::error::Error for #ty #ty_generics #error_where_clause {
             #source_method
             #provide_method
@@ -427,6 +432,7 @@ fn impl_enum(input: Enum) -> TokenStream {
         let display_where_clause = display_inferred_bounds.augment_where_clause(input.generics);
         Some(quote! {
             #[allow(unused_qualifications)]
+            #[automatically_derived]
             impl #impl_generics ::core::fmt::Display for #ty #ty_generics #display_where_clause {
                 fn fmt(&self, __formatter: &mut ::core::fmt::Formatter) -> ::core::fmt::Result {
                     #use_as_display
@@ -449,6 +455,7 @@ fn impl_enum(input: Enum) -> TokenStream {
         let body = from_initializer(from_field, backtrace_field);
         Some(quote! {
             #[allow(unused_qualifications)]
+            #[automatically_derived]
             impl #impl_generics ::core::convert::From<#from> for #ty #ty_generics #where_clause {
                 #[allow(deprecated)]
                 fn from(source: #from) -> Self {
@@ -467,6 +474,7 @@ fn impl_enum(input: Enum) -> TokenStream {
 
     quote! {
         #[allow(unused_qualifications)]
+        #[automatically_derived]
         impl #impl_generics std::error::Error for #ty #ty_generics #error_where_clause {
             #source_method
             #provide_method

third_party/rust/thiserror/Cargo.toml

@@ -13,7 +13,7 @@
 edition = "2021"
 rust-version = "1.56"
 name = "thiserror"
-version = "1.0.63"
+version = "1.0.64"
 authors = ["David Tolnay <dtolnay@gmail.com>"]
 build = "build.rs"
 autobins = false
@@ -93,7 +93,7 @@ name = "test_transparent"
 path = "tests/test_transparent.rs"
 
 [dependencies.thiserror-impl]
-version = "=1.0.63"
+version = "=1.0.64"
 
 [dev-dependencies.anyhow]
 version = "1.0.73"

third_party/rust/thiserror/src/lib.rs

@@ -258,7 +258,7 @@
 //!
 //!   [`anyhow`]: https://github.com/dtolnay/anyhow
 
-#![doc(html_root_url = "https://docs.rs/thiserror/1.0.63")]
+#![doc(html_root_url = "https://docs.rs/thiserror/1.0.64")]
 #![allow(
     clippy::module_name_repetitions,
     clippy::needless_lifetimes,

third_party/rust/thiserror/tests/test_generics.rs

@@ -158,4 +158,4 @@ pub struct StructFromGeneric<E> {
 //
 #[derive(Error, Debug)]
 #[error(transparent)]
-pub struct StructTransparentGeneric<E>(E);
+pub struct StructTransparentGeneric<E>(pub E);

third_party/rust/unicode-xid/.cargo-checksum.json

@@ -1 +1 @@
-{"files":{"COPYRIGHT":"23860c2a7b5d96b21569afedf033469bab9fe14a1b24a35068b8641c578ce24d","Cargo.toml":"03a1994da0569af5311597c8b43633512303e7fc3e8f93107b5b69478b249155","LICENSE-APACHE":"a60eea817514531668d7e00765731449fe14d059d3249e0bc93b36de45f759f2","LICENSE-MIT":"7b63ecd5f1902af1b63729947373683c32745c16a10e8e6292e2e2dcd7e90ae0","README.md":"5a4b2a4f0abc2877a7c6a7652b884539ca98bc88ccb2003d42fde89d1b635b93","benches/xid.rs":"a2986ce1df5d93bff51e73dc234bffb341d4fe5d749247296f02396dde16a72b","src/lib.rs":"5ec8dbe7a10106d8798e8469986408ecfb60710e7f6a936e265283753c8d1a94","src/tables.rs":"bb89d0f90360bb871577990f8ae0299b6ac8afcd4205be14b78947660a01455f","src/tests.rs":"56da0662966bb3c69201ae3eb1f69a62b808d604584b7fbdc4f930905301e4c6","tests/exhaustive_tests.rs":"ec91e9124d61e5b3e2fbbbf37a2fd30d6c7a8fa22639d6720b416d9ebc1007c5"},"package":"229730647fbc343e3a80e463c1db7f78f3855d3f3739bee0dda773c9a037c90a"}
+{"files":{"COPYRIGHT":"23860c2a7b5d96b21569afedf033469bab9fe14a1b24a35068b8641c578ce24d","Cargo.toml":"030ed8e60fcd73b831f174ab2df10e0f913d5d1476909c92aee26849323370f3","LICENSE-APACHE":"a60eea817514531668d7e00765731449fe14d059d3249e0bc93b36de45f759f2","LICENSE-MIT":"7b63ecd5f1902af1b63729947373683c32745c16a10e8e6292e2e2dcd7e90ae0","README.md":"651a08ce6a8967dfd5eaa7de96f527c59d3ca2cac2754b8a5ff7a014b7551700","benches/xid.rs":"a2986ce1df5d93bff51e73dc234bffb341d4fe5d749247296f02396dde16a72b","src/lib.rs":"5ec8dbe7a10106d8798e8469986408ecfb60710e7f6a936e265283753c8d1a94","src/tables.rs":"6a886d9107d42e8dbec81fa7f4acbb6a82adcec4976b56f757551c7a61cf788d","src/tests.rs":"56da0662966bb3c69201ae3eb1f69a62b808d604584b7fbdc4f930905301e4c6","tests/exhaustive_tests.rs":"ec91e9124d61e5b3e2fbbbf37a2fd30d6c7a8fa22639d6720b416d9ebc1007c5"},"package":"ebc1c04c71510c7f702b52b7c350734c9ff1295c464a03335b00bb84fc54f853"}

third_party/rust/unicode-xid/Cargo.toml

@@ -12,7 +12,7 @@
 [package]
 rust-version = "1.17"
 name = "unicode-xid"
-version = "0.2.5"
+version = "0.2.6"
 authors = [
     "erick.tryzelaar <erick.tryzelaar@gmail.com>",
     "kwantam <kwantam@gmail.com>",

third_party/rust/unicode-xid/README.md

@@ -26,6 +26,10 @@ on std, and instead uses equivalent functions from core.
 
 # changelog
 
+## 0.2.6
+
+- Update to Unicode 16.0.0.
+
 ## 0.2.5
 
 - Update to Unicode 15.1.0.

third_party/rust/unicode-xid/src/tables.rs

@@ -14,7 +14,7 @@
 
 /// The version of [Unicode](http://www.unicode.org/)
 /// that this version of unicode-xid is based on.
-pub const UNICODE_VERSION: (u64, u64, u64) = (15, 1, 0);
+pub const UNICODE_VERSION: (u64, u64, u64) = (16, 0, 0);
 
 fn bsearch_range_table(c: char, r: &[(char, char)]) -> bool {
     use core::cmp::Ordering::{Equal, Greater, Less};
@@ -91,7 +91,7 @@ pub mod derived_property {
         ('\u{860}', '\u{86a}'),
         ('\u{870}', '\u{887}'),
         ('\u{889}', '\u{88e}'),
-        ('\u{898}', '\u{8e1}'),
+        ('\u{897}', '\u{8e1}'),
         ('\u{8e3}', '\u{963}'),
         ('\u{966}', '\u{96f}'),
         ('\u{971}', '\u{983}'),
@@ -314,7 +314,7 @@ pub mod derived_property {
         ('\u{1c00}', '\u{1c37}'),
         ('\u{1c40}', '\u{1c49}'),
         ('\u{1c4d}', '\u{1c7d}'),
-        ('\u{1c80}', '\u{1c88}'),
+        ('\u{1c80}', '\u{1c8a}'),
         ('\u{1c90}', '\u{1cba}'),
         ('\u{1cbd}', '\u{1cbf}'),
         ('\u{1cd0}', '\u{1cd2}'),
@@ -399,10 +399,10 @@ pub mod derived_property {
         ('\u{a67f}', '\u{a6f1}'),
         ('\u{a717}', '\u{a71f}'),
         ('\u{a722}', '\u{a788}'),
-        ('\u{a78b}', '\u{a7ca}'),
+        ('\u{a78b}', '\u{a7cd}'),
         ('\u{a7d0}', '\u{a7d1}'),
         ('\u{a7d3}', '\u{a7d3}'),
-        ('\u{a7d5}', '\u{a7d9}'),
+        ('\u{a7d5}', '\u{a7dc}'),
         ('\u{a7f2}', '\u{a827}'),
         ('\u{a82c}', '\u{a82c}'),
         ('\u{a840}', '\u{a873}'),
@@ -506,6 +506,7 @@ pub mod derived_property {
         ('\u{105a3}', '\u{105b1}'),
         ('\u{105b3}', '\u{105b9}'),
         ('\u{105bb}', '\u{105bc}'),
+        ('\u{105c0}', '\u{105f3}'),
         ('\u{10600}', '\u{10736}'),
         ('\u{10740}', '\u{10755}'),
         ('\u{10760}', '\u{10767}'),
@@ -546,10 +547,14 @@ pub mod derived_property {
         ('\u{10cc0}', '\u{10cf2}'),
         ('\u{10d00}', '\u{10d27}'),
         ('\u{10d30}', '\u{10d39}'),
+        ('\u{10d40}', '\u{10d65}'),
+        ('\u{10d69}', '\u{10d6d}'),
+        ('\u{10d6f}', '\u{10d85}'),
         ('\u{10e80}', '\u{10ea9}'),
         ('\u{10eab}', '\u{10eac}'),
         ('\u{10eb0}', '\u{10eb1}'),
-        ('\u{10efd}', '\u{10f1c}'),
+        ('\u{10ec2}', '\u{10ec4}'),
+        ('\u{10efc}', '\u{10f1c}'),
         ('\u{10f27}', '\u{10f27}'),
         ('\u{10f30}', '\u{10f50}'),
         ('\u{10f70}', '\u{10f85}'),
@@ -595,6 +600,16 @@ pub mod derived_property {
         ('\u{1135d}', '\u{11363}'),
         ('\u{11366}', '\u{1136c}'),
         ('\u{11370}', '\u{11374}'),
+        ('\u{11380}', '\u{11389}'),
+        ('\u{1138b}', '\u{1138b}'),
+        ('\u{1138e}', '\u{1138e}'),
+        ('\u{11390}', '\u{113b5}'),
+        ('\u{113b7}', '\u{113c0}'),
+        ('\u{113c2}', '\u{113c2}'),
+        ('\u{113c5}', '\u{113c5}'),
+        ('\u{113c7}', '\u{113ca}'),
+        ('\u{113cc}', '\u{113d3}'),
+        ('\u{113e1}', '\u{113e2}'),
         ('\u{11400}', '\u{1144a}'),
         ('\u{11450}', '\u{11459}'),
         ('\u{1145e}', '\u{11461}'),
@@ -609,6 +624,7 @@ pub mod derived_property {
         ('\u{11650}', '\u{11659}'),
         ('\u{11680}', '\u{116b8}'),
         ('\u{116c0}', '\u{116c9}'),
+        ('\u{116d0}', '\u{116e3}'),
         ('\u{11700}', '\u{1171a}'),
         ('\u{1171d}', '\u{1172b}'),
         ('\u{11730}', '\u{11739}'),
@@ -632,6 +648,8 @@ pub mod derived_property {
         ('\u{11a50}', '\u{11a99}'),
         ('\u{11a9d}', '\u{11a9d}'),
         ('\u{11ab0}', '\u{11af8}'),
+        ('\u{11bc0}', '\u{11be0}'),
+        ('\u{11bf0}', '\u{11bf9}'),
         ('\u{11c00}', '\u{11c08}'),
         ('\u{11c0a}', '\u{11c36}'),
         ('\u{11c38}', '\u{11c40}'),
@@ -656,7 +674,7 @@ pub mod derived_property {
         ('\u{11f00}', '\u{11f10}'),
         ('\u{11f12}', '\u{11f3a}'),
         ('\u{11f3e}', '\u{11f42}'),
-        ('\u{11f50}', '\u{11f59}'),
+        ('\u{11f50}', '\u{11f5a}'),
         ('\u{11fb0}', '\u{11fb0}'),
         ('\u{12000}', '\u{12399}'),
         ('\u{12400}', '\u{1246e}'),
@@ -664,7 +682,9 @@ pub mod derived_property {
         ('\u{12f90}', '\u{12ff0}'),
         ('\u{13000}', '\u{1342f}'),
         ('\u{13440}', '\u{13455}'),
+        ('\u{13460}', '\u{143fa}'),
         ('\u{14400}', '\u{14646}'),
+        ('\u{16100}', '\u{16139}'),
         ('\u{16800}', '\u{16a38}'),
         ('\u{16a40}', '\u{16a5e}'),
         ('\u{16a60}', '\u{16a69}'),
@@ -677,6 +697,8 @@ pub mod derived_property {
         ('\u{16b50}', '\u{16b59}'),
         ('\u{16b63}', '\u{16b77}'),
         ('\u{16b7d}', '\u{16b8f}'),
+        ('\u{16d40}', '\u{16d6c}'),
+        ('\u{16d70}', '\u{16d79}'),
         ('\u{16e40}', '\u{16e7f}'),
         ('\u{16f00}', '\u{16f4a}'),
         ('\u{16f4f}', '\u{16f87}'),
@@ -686,7 +708,7 @@ pub mod derived_property {
         ('\u{16ff0}', '\u{16ff1}'),
         ('\u{17000}', '\u{187f7}'),
         ('\u{18800}', '\u{18cd5}'),
-        ('\u{18d00}', '\u{18d08}'),
+        ('\u{18cff}', '\u{18d08}'),
         ('\u{1aff0}', '\u{1aff3}'),
         ('\u{1aff5}', '\u{1affb}'),
         ('\u{1affd}', '\u{1affe}'),
@@ -701,6 +723,7 @@ pub mod derived_property {
         ('\u{1bc80}', '\u{1bc88}'),
         ('\u{1bc90}', '\u{1bc99}'),
         ('\u{1bc9d}', '\u{1bc9e}'),
+        ('\u{1ccf0}', '\u{1ccf9}'),
         ('\u{1cf00}', '\u{1cf2d}'),
         ('\u{1cf30}', '\u{1cf46}'),
         ('\u{1d165}', '\u{1d169}'),
@@ -762,6 +785,7 @@ pub mod derived_property {
         ('\u{1e290}', '\u{1e2ae}'),
         ('\u{1e2c0}', '\u{1e2f9}'),
         ('\u{1e4d0}', '\u{1e4f9}'),
+        ('\u{1e5d0}', '\u{1e5fa}'),
         ('\u{1e7e0}', '\u{1e7e6}'),
         ('\u{1e7e8}', '\u{1e7eb}'),
         ('\u{1e7ed}', '\u{1e7ee}'),
@@ -1045,7 +1069,7 @@ pub mod derived_property {
         ('\u{1c00}', '\u{1c23}'),
         ('\u{1c4d}', '\u{1c4f}'),
         ('\u{1c5a}', '\u{1c7d}'),
-        ('\u{1c80}', '\u{1c88}'),
+        ('\u{1c80}', '\u{1c8a}'),
         ('\u{1c90}', '\u{1cba}'),
         ('\u{1cbd}', '\u{1cbf}'),
         ('\u{1ce9}', '\u{1cec}'),
@@ -1128,10 +1152,10 @@ pub mod derived_property {
         ('\u{a6a0}', '\u{a6ef}'),
         ('\u{a717}', '\u{a71f}'),
         ('\u{a722}', '\u{a788}'),
-        ('\u{a78b}', '\u{a7ca}'),
+        ('\u{a78b}', '\u{a7cd}'),
         ('\u{a7d0}', '\u{a7d1}'),
         ('\u{a7d3}', '\u{a7d3}'),
-        ('\u{a7d5}', '\u{a7d9}'),
+        ('\u{a7d5}', '\u{a7dc}'),
         ('\u{a7f2}', '\u{a801}'),
         ('\u{a803}', '\u{a805}'),
         ('\u{a807}', '\u{a80a}'),
@@ -1236,6 +1260,7 @@ pub mod derived_property {
         ('\u{105a3}', '\u{105b1}'),
         ('\u{105b3}', '\u{105b9}'),
         ('\u{105bb}', '\u{105bc}'),
+        ('\u{105c0}', '\u{105f3}'),
         ('\u{10600}', '\u{10736}'),
         ('\u{10740}', '\u{10755}'),
         ('\u{10760}', '\u{10767}'),
@@ -1272,8 +1297,11 @@ pub mod derived_property {
         ('\u{10c80}', '\u{10cb2}'),
         ('\u{10cc0}', '\u{10cf2}'),
         ('\u{10d00}', '\u{10d23}'),
+        ('\u{10d4a}', '\u{10d65}'),
+        ('\u{10d6f}', '\u{10d85}'),
         ('\u{10e80}', '\u{10ea9}'),
         ('\u{10eb0}', '\u{10eb1}'),
+        ('\u{10ec2}', '\u{10ec4}'),
         ('\u{10f00}', '\u{10f1c}'),
         ('\u{10f27}', '\u{10f27}'),
         ('\u{10f30}', '\u{10f45}'),
@@ -1312,6 +1340,13 @@ pub mod derived_property {
         ('\u{1133d}', '\u{1133d}'),
         ('\u{11350}', '\u{11350}'),
         ('\u{1135d}', '\u{11361}'),
+        ('\u{11380}', '\u{11389}'),
+        ('\u{1138b}', '\u{1138b}'),
+        ('\u{1138e}', '\u{1138e}'),
+        ('\u{11390}', '\u{113b5}'),
+        ('\u{113b7}', '\u{113b7}'),
+        ('\u{113d1}', '\u{113d1}'),
+        ('\u{113d3}', '\u{113d3}'),
         ('\u{11400}', '\u{11434}'),
         ('\u{11447}', '\u{1144a}'),
         ('\u{1145f}', '\u{11461}'),
@@ -1346,6 +1381,7 @@ pub mod derived_property {
         ('\u{11a5c}', '\u{11a89}'),
         ('\u{11a9d}', '\u{11a9d}'),
         ('\u{11ab0}', '\u{11af8}'),
+        ('\u{11bc0}', '\u{11be0}'),
         ('\u{11c00}', '\u{11c08}'),
         ('\u{11c0a}', '\u{11c2e}'),
         ('\u{11c40}', '\u{11c40}'),
@@ -1369,7 +1405,9 @@ pub mod derived_property {
         ('\u{12f90}', '\u{12ff0}'),
         ('\u{13000}', '\u{1342f}'),
         ('\u{13441}', '\u{13446}'),
+        ('\u{13460}', '\u{143fa}'),
         ('\u{14400}', '\u{14646}'),
+        ('\u{16100}', '\u{1611d}'),
         ('\u{16800}', '\u{16a38}'),
         ('\u{16a40}', '\u{16a5e}'),
         ('\u{16a70}', '\u{16abe}'),
@@ -1378,6 +1416,7 @@ pub mod derived_property {
         ('\u{16b40}', '\u{16b43}'),
         ('\u{16b63}', '\u{16b77}'),
         ('\u{16b7d}', '\u{16b8f}'),
+        ('\u{16d40}', '\u{16d6c}'),
         ('\u{16e40}', '\u{16e7f}'),
         ('\u{16f00}', '\u{16f4a}'),
         ('\u{16f50}', '\u{16f50}'),
@@ -1386,7 +1425,7 @@ pub mod derived_property {
         ('\u{16fe3}', '\u{16fe3}'),
         ('\u{17000}', '\u{187f7}'),
         ('\u{18800}', '\u{18cd5}'),
-        ('\u{18d00}', '\u{18d08}'),
+        ('\u{18cff}', '\u{18d08}'),
         ('\u{1aff0}', '\u{1aff3}'),
         ('\u{1aff5}', '\u{1affb}'),
         ('\u{1affd}', '\u{1affe}'),
@@ -1439,6 +1478,8 @@ pub mod derived_property {
         ('\u{1e290}', '\u{1e2ad}'),
         ('\u{1e2c0}', '\u{1e2eb}'),
         ('\u{1e4d0}', '\u{1e4eb}'),
+        ('\u{1e5d0}', '\u{1e5ed}'),
+        ('\u{1e5f0}', '\u{1e5f0}'),
         ('\u{1e7e0}', '\u{1e7e6}'),
         ('\u{1e7e8}', '\u{1e7eb}'),
         ('\u{1e7ed}', '\u{1e7ee}'),

third_party/rust/wgpu-core/Cargo.toml

@@ -58,7 +58,6 @@ version = "2.6"
 
 [dependencies.bytemuck]
 version = "1.18"
-features = ["derive"]
 optional = true
 
 [dependencies.document-features]
@@ -110,7 +109,7 @@ optional = true
 version = "1"
 
 [dependencies.thiserror]
-version = "1.0.63"
+version = "1.0.64"
 
 [dependencies.wgt]
 version = "22.0.0"

third_party/rust/wgpu-core/src/binding_model.rs

@@ -373,10 +373,6 @@ impl BindingTypeMaxCountValidator {
             limits.max_sampled_textures_per_shader_stage,
             BindingTypeMaxCountErrorKind::SampledTextures,
         )?;
-        self.storage_buffers.validate(
-            limits.max_storage_buffers_per_shader_stage,
-            BindingTypeMaxCountErrorKind::StorageBuffers,
-        )?;
         self.samplers.validate(
             limits.max_samplers_per_shader_stage,
             BindingTypeMaxCountErrorKind::Samplers,

third_party/rust/wgpu-core/src/command/bundle.rs

@@ -1584,8 +1584,7 @@ pub mod bundle_ffi {
     ///
     /// This function is unsafe as there is no guarantee that the given pointer is
     /// valid for `offset_length` elements.
-    #[no_mangle]
-    pub unsafe extern "C" fn wgpu_render_bundle_set_bind_group(
+    pub unsafe fn wgpu_render_bundle_set_bind_group(
         bundle: &mut RenderBundleEncoder,
         index: u32,
         bind_group_id: Option<id::BindGroupId>,
@@ -1612,8 +1611,7 @@ pub mod bundle_ffi {
         });
     }
 
-    #[no_mangle]
-    pub extern "C" fn wgpu_render_bundle_set_pipeline(
+    pub fn wgpu_render_bundle_set_pipeline(
         bundle: &mut RenderBundleEncoder,
         pipeline_id: id::RenderPipelineId,
     ) {
@@ -1627,8 +1625,7 @@ pub mod bundle_ffi {
             .push(RenderCommand::SetPipeline(pipeline_id));
     }
 
-    #[no_mangle]
-    pub extern "C" fn wgpu_render_bundle_set_vertex_buffer(
+    pub fn wgpu_render_bundle_set_vertex_buffer(
         bundle: &mut RenderBundleEncoder,
         slot: u32,
         buffer_id: id::BufferId,
@@ -1643,8 +1640,7 @@ pub mod bundle_ffi {
         });
     }
 
-    #[no_mangle]
-    pub extern "C" fn wgpu_render_bundle_set_index_buffer(
+    pub fn wgpu_render_bundle_set_index_buffer(
         encoder: &mut RenderBundleEncoder,
         buffer: id::BufferId,
         index_format: IndexFormat,
@@ -1658,8 +1654,7 @@ pub mod bundle_ffi {
     ///
     /// This function is unsafe as there is no guarantee that the given pointer is
     /// valid for `data` elements.
-    #[no_mangle]
-    pub unsafe extern "C" fn wgpu_render_bundle_set_push_constants(
+    pub unsafe fn wgpu_render_bundle_set_push_constants(
         pass: &mut RenderBundleEncoder,
         stages: wgt::ShaderStages,
         offset: u32,
@@ -1695,8 +1690,7 @@ pub mod bundle_ffi {
         });
     }
 
-    #[no_mangle]
-    pub extern "C" fn wgpu_render_bundle_draw(
+    pub fn wgpu_render_bundle_draw(
         bundle: &mut RenderBundleEncoder,
         vertex_count: u32,
         instance_count: u32,
@@ -1711,8 +1705,7 @@ pub mod bundle_ffi {
         });
     }
 
-    #[no_mangle]
-    pub extern "C" fn wgpu_render_bundle_draw_indexed(
+    pub fn wgpu_render_bundle_draw_indexed(
         bundle: &mut RenderBundleEncoder,
         index_count: u32,
         instance_count: u32,
@@ -1729,8 +1722,7 @@ pub mod bundle_ffi {
         });
     }
 
-    #[no_mangle]
-    pub extern "C" fn wgpu_render_bundle_draw_indirect(
+    pub fn wgpu_render_bundle_draw_indirect(
         bundle: &mut RenderBundleEncoder,
         buffer_id: id::BufferId,
         offset: BufferAddress,
@@ -1743,8 +1735,7 @@ pub mod bundle_ffi {
         });
     }
 
-    #[no_mangle]
-    pub extern "C" fn wgpu_render_bundle_draw_indexed_indirect(
+    pub fn wgpu_render_bundle_draw_indexed_indirect(
         bundle: &mut RenderBundleEncoder,
         buffer_id: id::BufferId,
         offset: BufferAddress,
@@ -1761,16 +1752,14 @@ pub mod bundle_ffi {
     ///
     /// This function is unsafe as there is no guarantee that the given `label`
     /// is a valid null-terminated string.
-    #[no_mangle]
-    pub unsafe extern "C" fn wgpu_render_bundle_push_debug_group(
+    pub unsafe fn wgpu_render_bundle_push_debug_group(
         _bundle: &mut RenderBundleEncoder,
         _label: RawString,
     ) {
         //TODO
     }
 
-    #[no_mangle]
-    pub extern "C" fn wgpu_render_bundle_pop_debug_group(_bundle: &mut RenderBundleEncoder) {
+    pub fn wgpu_render_bundle_pop_debug_group(_bundle: &mut RenderBundleEncoder) {
         //TODO
     }
 
@@ -1778,8 +1767,7 @@ pub mod bundle_ffi {
     ///
     /// This function is unsafe as there is no guarantee that the given `label`
     /// is a valid null-terminated string.
-    #[no_mangle]
-    pub unsafe extern "C" fn wgpu_render_bundle_insert_debug_marker(
+    pub unsafe fn wgpu_render_bundle_insert_debug_marker(
         _bundle: &mut RenderBundleEncoder,
         _label: RawString,
     ) {

third_party/rust/wgpu-core/src/command/compute.rs

@@ -136,6 +136,8 @@ pub enum ComputePassErrorInner {
     BindGroupIndexOutOfRange { index: u32, max: u32 },
     #[error(transparent)]
     DestroyedResource(#[from] DestroyedResourceError),
+    #[error("Indirect buffer offset {0:?} is not a multiple of 4")]
+    UnalignedIndirectBufferOffset(BufferAddress),
     #[error("Indirect buffer uses bytes {offset}..{end_offset} which overruns indirect buffer of size {buffer_size}")]
     IndirectBufferOverrun {
         offset: u64,
@@ -495,8 +497,6 @@ impl Global {
             state.raw_encoder.begin_compute_pass(&hal_desc);
         }
 
-        // TODO: We should be draining the commands here, avoiding extra copies in the process.
-        //       (A command encoder can't be executed twice!)
         for command in base.commands {
             match command {
                 ArcComputeCommand::SetBindGroup {
@@ -780,7 +780,7 @@ fn set_push_constant(
         .binder
         .pipeline_layout
         .as_ref()
-        //TODO: don't error here, lazily update the push constants
+        // TODO: don't error here, lazily update the push constants using `state.push_constants`
         .ok_or(ComputePassErrorInner::Dispatch(
             DispatchError::MissingPipeline,
         ))?;
@@ -847,6 +847,10 @@ fn dispatch_indirect(
         .merge_single(&buffer, hal::BufferUses::INDIRECT)?;
     buffer.check_usage(wgt::BufferUsages::INDIRECT)?;
 
+    if offset % 4 != 0 {
+        return Err(ComputePassErrorInner::UnalignedIndirectBufferOffset(offset));
+    }
+
     let end_offset = offset + size_of::<wgt::DispatchIndirectArgs>() as u64;
     if end_offset > buffer.size {
         return Err(ComputePassErrorInner::IndirectBufferOverrun {

third_party/rust/wgpu-core/src/command/render.rs

@@ -638,6 +638,8 @@ pub enum RenderPassErrorInner {
     MissingFeatures(#[from] MissingFeatures),
     #[error(transparent)]
     MissingDownlevelFlags(#[from] MissingDownlevelFlags),
+    #[error("Indirect buffer offset {0:?} is not a multiple of 4")]
+    UnalignedIndirectBufferOffset(BufferAddress),
     #[error("Indirect draw uses bytes {offset}..{end_offset} {} which overruns indirect buffer of size {buffer_size}",
         count.map_or_else(String::new, |v| format!("(using count {v})")))]
     IndirectBufferOverrun {
@@ -2450,6 +2452,10 @@ fn multi_draw_indirect(
 
     let actual_count = count.map_or(1, |c| c.get());
 
+    if offset % 4 != 0 {
+        return Err(RenderPassErrorInner::UnalignedIndirectBufferOffset(offset));
+    }
+
     let end_offset = offset + stride as u64 * actual_count as u64;
     if end_offset > indirect_buffer.size {
         return Err(RenderPassErrorInner::IndirectBufferOverrun {
@@ -2534,6 +2540,10 @@ fn multi_draw_indirect_count(
     count_buffer.check_usage(BufferUsages::INDIRECT)?;
     let count_raw = count_buffer.try_raw(state.snatch_guard)?;
 
+    if offset % 4 != 0 {
+        return Err(RenderPassErrorInner::UnalignedIndirectBufferOffset(offset));
+    }
+
     let end_offset = offset + stride * max_count as u64;
     if end_offset > indirect_buffer.size {
         return Err(RenderPassErrorInner::IndirectBufferOverrun {

third_party/rust/wgpu-core/src/device/queue.rs

@@ -1027,11 +1027,13 @@ impl Global {
         &self,
         queue_id: QueueId,
         command_buffer_ids: &[id::CommandBufferId],
-    ) -> Result<SubmissionIndex, QueueSubmitError> {
+    ) -> Result<SubmissionIndex, (SubmissionIndex, QueueSubmitError)> {
         profiling::scope!("Queue::submit");
         api_log!("Queue::submit {queue_id:?}");
 
-        let (submit_index, callbacks) = {
+        let submit_index;
+
+        let res = 'error: {
             let hub = &self.hub;
 
             let queue = hub.queues.get(queue_id);
@@ -1042,7 +1044,7 @@ impl Global {
 
             // Fence lock must be acquired after the snatch lock everywhere to avoid deadlocks.
             let mut fence = device.fence.write();
-            let submit_index = device
+            submit_index = device
                 .active_submission_index
                 .fetch_add(1, Ordering::SeqCst)
                 + 1;
@@ -1119,18 +1121,29 @@ impl Global {
                         }
 
                         // execute resource transitions
-                        unsafe {
+                        if let Err(e) = unsafe {
                             baked.encoder.begin_encoding(hal_label(
                                 Some("(wgpu internal) Transit"),
                                 device.instance_flags,
                             ))
                         }
-                        .map_err(|e| device.handle_hal_error(e))?;
+                        .map_err(|e| device.handle_hal_error(e))
+                        {
+                            break 'error Err(e.into());
+                        }
 
                         //Note: locking the trackers has to be done after the storages
                         let mut trackers = device.trackers.lock();
-                        baked.initialize_buffer_memory(&mut trackers, &snatch_guard)?;
-                        baked.initialize_texture_memory(&mut trackers, device, &snatch_guard)?;
+                        if let Err(e) = baked.initialize_buffer_memory(&mut trackers, &snatch_guard)
+                        {
+                            break 'error Err(e.into());
+                        }
+                        if let Err(e) =
+                            baked.initialize_texture_memory(&mut trackers, device, &snatch_guard)
+                        {
+                            break 'error Err(e.into());
+                        }
+
                         //Note: stateless trackers are not merged:
                         // device already knows these resources exist.
                         CommandBuffer::insert_barriers_from_device_tracker(
@@ -1147,13 +1160,16 @@ impl Global {
                         // Note: we could technically do it after all of the command buffers,
                         // but here we have a command encoder by hand, so it's easier to use it.
                         if !used_surface_textures.is_empty() {
-                            unsafe {
+                            if let Err(e) = unsafe {
                                 baked.encoder.begin_encoding(hal_label(
                                     Some("(wgpu internal) Present"),
                                     device.instance_flags,
                                 ))
                             }
-                            .map_err(|e| device.handle_hal_error(e))?;
+                            .map_err(|e| device.handle_hal_error(e))
+                            {
+                                break 'error Err(e.into());
+                            }
                             let texture_barriers = trackers
                                 .textures
                                 .set_from_usage_scope_and_drain_transitions(
@@ -1180,7 +1196,7 @@ impl Global {
                     }
 
                     if let Some(first_error) = first_error {
-                        return Err(first_error);
+                        break 'error Err(first_error);
                     }
                 }
             }
@@ -1190,9 +1206,9 @@ impl Global {
             {
                 used_surface_textures.set_size(hub.textures.read().len());
                 for texture in pending_writes.dst_textures.values() {
-                    match texture.try_inner(&snatch_guard)? {
-                        TextureInner::Native { .. } => {}
-                        TextureInner::Surface { .. } => {
+                    match texture.try_inner(&snatch_guard) {
+                        Ok(TextureInner::Native { .. }) => {}
+                        Ok(TextureInner::Surface { .. }) => {
                             // Compare the Arcs by pointer as Textures don't implement Eq
                             submit_surface_textures_owned
                                 .insert(Arc::as_ptr(texture), texture.clone());
@@ -1203,6 +1219,7 @@ impl Global {
                                     .unwrap()
                             };
                         }
+                        Err(e) => break 'error Err(e.into()),
                     }
                 }
 
@@ -1224,10 +1241,12 @@ impl Global {
                 }
             }
 
-            if let Some(pending_execution) =
-                pending_writes.pre_submit(&device.command_allocator, device, &queue)?
-            {
-                active_executions.insert(0, pending_execution);
+            match pending_writes.pre_submit(&device.command_allocator, device, &queue) {
+                Ok(Some(pending_execution)) => {
+                    active_executions.insert(0, pending_execution);
+                }
+                Ok(None) => {}
+                Err(e) => break 'error Err(e.into()),
             }
 
             let hal_command_buffers = active_executions
@@ -1249,14 +1268,17 @@ impl Global {
                     submit_surface_textures.push(raw);
                 }
 
-                unsafe {
+                if let Err(e) = unsafe {
                     queue.raw().submit(
                         &hal_command_buffers,
                         &submit_surface_textures,
                         (fence.as_mut(), submit_index),
                     )
                 }
-                .map_err(|e| device.handle_hal_error(e))?;
+                .map_err(|e| device.handle_hal_error(e))
+                {
+                    break 'error Err(e.into());
+                }
 
                 // Advance the successful submission index.
                 device
@@ -1280,12 +1302,19 @@ impl Global {
             let (closures, _) =
                 match device.maintain(fence_guard, wgt::Maintain::Poll, snatch_guard) {
                     Ok(closures) => closures,
-                    Err(WaitIdleError::Device(err)) => return Err(QueueSubmitError::Queue(err)),
-                    Err(WaitIdleError::StuckGpu) => return Err(QueueSubmitError::StuckGpu),
+                    Err(WaitIdleError::Device(err)) => {
+                        break 'error Err(QueueSubmitError::Queue(err))
+                    }
+                    Err(WaitIdleError::StuckGpu) => break 'error Err(QueueSubmitError::StuckGpu),
                     Err(WaitIdleError::WrongSubmissionIndex(..)) => unreachable!(),
                 };
 
-            (submit_index, closures)
+            Ok(closures)
+        };
+
+        let callbacks = match res {
+            Ok(ok) => ok,
+            Err(e) => return Err((submit_index, e)),
         };
 
         // the closures should execute with nothing locked!

third_party/rust/wgpu-core/src/device/resource.rs

@@ -38,7 +38,6 @@ use arrayvec::ArrayVec;
 use once_cell::sync::OnceCell;
 
 use smallvec::SmallVec;
-use thiserror::Error;
 use wgt::{
     math::align_to, DeviceLostReason, TextureFormat, TextureSampleType, TextureViewDimension,
 };
@@ -187,14 +186,6 @@ impl Drop for Device {
     }
 }
 
-#[derive(Clone, Debug, Error)]
-pub enum CreateDeviceError {
-    #[error("Not enough memory left to create device")]
-    OutOfMemory,
-    #[error("Failed to create internal buffer for initializing textures")]
-    FailedToCreateZeroBuffer(#[from] DeviceError),
-}
-
 impl Device {
     pub(crate) fn raw(&self) -> &dyn hal::DynDevice {
         self.raw.as_ref()
@@ -372,7 +363,7 @@ impl Device {
                     let Some(view) = view.upgrade() else {
                         continue;
                     };
-                    let Some(raw_view) = view.raw.snatch(self.snatchable_lock.write()) else {
+                    let Some(raw_view) = view.raw.snatch(&mut self.snatchable_lock.write()) else {
                         continue;
                     };
 
@@ -386,7 +377,8 @@ impl Device {
                     let Some(bind_group) = bind_group.upgrade() else {
                         continue;
                     };
-                    let Some(raw_bind_group) = bind_group.raw.snatch(self.snatchable_lock.write())
+                    let Some(raw_bind_group) =
+                        bind_group.raw.snatch(&mut self.snatchable_lock.write())
                     else {
                         continue;
                     };
@@ -437,13 +429,11 @@ impl Device {
                     .last_successful_submission_index
                     .load(Ordering::Acquire);
 
-                if let wgt::Maintain::WaitForSubmissionIndex(submission_index) = maintain {
-                    if submission_index > last_successful_submission_index {
-                        return Err(WaitIdleError::WrongSubmissionIndex(
-                            submission_index,
-                            last_successful_submission_index,
-                        ));
-                    }
+                if submission_index > last_successful_submission_index {
+                    return Err(WaitIdleError::WrongSubmissionIndex(
+                        submission_index,
+                        last_successful_submission_index,
+                    ));
                 }
 
                 submission_index
@@ -457,13 +447,13 @@ impl Device {
 
         // If necessary, wait for that submission to complete.
         if maintain.is_wait() {
+            log::trace!("Device::maintain: waiting for submission index {submission_index}");
             unsafe {
                 self.raw()
                     .wait(fence.as_ref(), submission_index, CLEANUP_WAIT_MS)
             }
             .map_err(|e| self.handle_hal_error(e))?;
         }
-        log::trace!("Device::maintain: waiting for submission index {submission_index}");
 
         let mut life_tracker = self.lock_life();
         let submission_closures =

third_party/rust/wgpu-core/src/id.rs

@@ -4,22 +4,23 @@ use std::{
     fmt::{self, Debug},
     hash::Hash,
     marker::PhantomData,
+    mem::size_of,
     num::NonZeroU64,
 };
 use wgt::WasmNotSendSync;
 
 const _: () = {
-    if std::mem::size_of::<Index>() != 4 {
+    if size_of::<Index>() != 4 {
         panic!()
     }
 };
 const _: () = {
-    if std::mem::size_of::<Epoch>() != 4 {
+    if size_of::<Epoch>() != 4 {
         panic!()
     }
 };
 const _: () = {
-    if std::mem::size_of::<RawId>() != 8 {
+    if size_of::<RawId>() != 8 {
         panic!()
     }
 };

third_party/rust/wgpu-core/src/lib.rs

@@ -128,16 +128,26 @@ pub fn hal_label(opt: Option<&str>, flags: wgt::InstanceFlags) -> Option<&str> {
     opt
 }
 
-const DOWNLEVEL_WARNING_MESSAGE: &str = "The underlying API or device in use does not \
-support enough features to be a fully compliant implementation of WebGPU. A subset of the features can still be used. \
-If you are running this program on native and not in a browser and wish to limit the features you use to the supported subset, \
-call Adapter::downlevel_properties or Device::downlevel_properties to get a listing of the features the current \
-platform supports.";
-const DOWNLEVEL_ERROR_MESSAGE: &str = "This is not an invalid use of WebGPU: the underlying API or device does not \
-support enough features to be a fully compliant implementation. A subset of the features can still be used. \
-If you are running this program on native and not in a browser and wish to work around this issue, call \
-Adapter::downlevel_properties or Device::downlevel_properties to get a listing of the features the current \
-platform supports.";
+const DOWNLEVEL_WARNING_MESSAGE: &str = concat!(
+    "The underlying API or device in use does not ",
+    "support enough features to be a fully compliant implementation of WebGPU. ",
+    "A subset of the features can still be used. ",
+    "If you are running this program on native and not in a browser and wish to limit ",
+    "the features you use to the supported subset, ",
+    "call Adapter::downlevel_properties or Device::downlevel_properties to get ",
+    "a listing of the features the current ",
+    "platform supports."
+);
+
+const DOWNLEVEL_ERROR_MESSAGE: &str = concat!(
+    "This is not an invalid use of WebGPU: the underlying API or device does not ",
+    "support enough features to be a fully compliant implementation. ",
+    "A subset of the features can still be used. ",
+    "If you are running this program on native and not in a browser ",
+    "and wish to work around this issue, call ",
+    "Adapter::downlevel_properties or Device::downlevel_properties ",
+    "to get a listing of the features the current platform supports."
+);
 
 #[cfg(feature = "api_log_info")]
 macro_rules! api_log {

third_party/rust/wgpu-core/src/present.rs

@@ -288,8 +288,11 @@ impl Global {
                     .textures
                     .remove(texture.tracker_index());
                 let suf = surface.raw(device.backend()).unwrap();
-                let exclusive_snatch_guard = device.snatchable_lock.write();
-                match texture.inner.snatch(exclusive_snatch_guard).unwrap() {
+                match texture
+                    .inner
+                    .snatch(&mut device.snatchable_lock.write())
+                    .unwrap()
+                {
                     resource::TextureInner::Surface { raw, parent_id } => {
                         if surface_id != parent_id {
                             log::error!("Presented frame is from a different surface");
@@ -359,8 +362,11 @@ impl Global {
                     .textures
                     .remove(texture.tracker_index());
                 let suf = surface.raw(device.backend());
-                let exclusive_snatch_guard = device.snatchable_lock.write();
-                match texture.inner.snatch(exclusive_snatch_guard).unwrap() {
+                match texture
+                    .inner
+                    .snatch(&mut device.snatchable_lock.write())
+                    .unwrap()
+                {
                     resource::TextureInner::Surface { raw, parent_id } => {
                         if surface_id == parent_id {
                             unsafe { suf.unwrap().discard_texture(raw) };

third_party/rust/wgpu-core/src/resource.rs

@@ -737,8 +737,7 @@ impl Buffer {
         let device = &self.device;
 
         let temp = {
-            let snatch_guard = device.snatchable_lock.write();
-            let raw = match self.raw.snatch(snatch_guard) {
+            let raw = match self.raw.snatch(&mut device.snatchable_lock.write()) {
                 Some(raw) => raw,
                 None => {
                     return Err(DestroyError::AlreadyDestroyed);
@@ -1185,8 +1184,7 @@ impl Texture {
         let device = &self.device;
 
         let temp = {
-            let snatch_guard = device.snatchable_lock.write();
-            let raw = match self.inner.snatch(snatch_guard) {
+            let raw = match self.inner.snatch(&mut device.snatchable_lock.write()) {
                 Some(TextureInner::Native { raw }) => raw,
                 Some(TextureInner::Surface { .. }) => {
                     return Ok(());

third_party/rust/wgpu-core/src/snatch.rs

@@ -38,7 +38,7 @@ impl<T> Snatchable<T> {
     }
 
     /// Take the value. Requires a the snatchable lock's write guard.
-    pub fn snatch(&self, _guard: ExclusiveSnatchGuard) -> Option<T> {
+    pub fn snatch(&self, _guard: &mut ExclusiveSnatchGuard) -> Option<T> {
         unsafe { (*self.value.get()).take() }
     }
 

third_party/rust/wgpu-hal/Cargo.toml

@@ -70,6 +70,10 @@ version = "0.7"
 [dependencies.bitflags]
 version = "2.6"
 
+[dependencies.bytemuck]
+version = "1.18"
+optional = true
+
 [dependencies.glow]
 version = "0.14.1"
 optional = true
@@ -98,7 +102,7 @@ version = "0.6"
 version = "1.1.0"
 
 [dependencies.thiserror]
-version = "1.0.63"
+version = "1.0.64"
 
 [dependencies.wgt]
 version = "22.0.0"
@@ -151,6 +155,7 @@ dx12 = [
 fragile-send-sync-non-atomic-wasm = ["wgt/fragile-send-sync-non-atomic-wasm"]
 gles = [
     "naga/glsl-out",
+    "dep:bytemuck",
     "dep:glow",
     "dep:glutin_wgl_sys",
     "dep:khronos-egl",

third_party/rust/wgpu-hal/examples/ray-traced-triangle/main.rs

@@ -124,7 +124,12 @@ impl AccelerationStructureInstance {
         &mut self,
         shader_binding_table_record_offset: u32,
     ) {
-        debug_assert!(shader_binding_table_record_offset <= Self::MAX_U24, "shader_binding_table_record_offset uses more than 24 bits! {shader_binding_table_record_offset} > {}", Self::MAX_U24);
+        debug_assert!(
+            shader_binding_table_record_offset <= Self::MAX_U24,
+            "shader_binding_table_record_offset uses more than 24 bits! {} > {}",
+            shader_binding_table_record_offset,
+            Self::MAX_U24
+        );
         self.shader_binding_table_record_offset_and_flags = (shader_binding_table_record_offset
             & Self::LOW_24_MASK)
             | (self.shader_binding_table_record_offset_and_flags & !Self::LOW_24_MASK)
@@ -151,7 +156,9 @@ impl AccelerationStructureInstance {
         );
         debug_assert!(
             shader_binding_table_record_offset <= Self::MAX_U24,
-            "shader_binding_table_record_offset uses more than 24 bits! {shader_binding_table_record_offset} > {}", Self::MAX_U24
+            "shader_binding_table_record_offset uses more than 24 bits! {} > {}",
+            shader_binding_table_record_offset,
+            Self::MAX_U24
         );
         AccelerationStructureInstance {
             transform: Self::affine_to_rows(transform),
@@ -205,7 +212,7 @@ struct Example<A: hal::Api> {
     uniform_buffer: A::Buffer,
     pipeline_layout: A::PipelineLayout,
     vertices_buffer: A::Buffer,
-    indices_buffer: A::Buffer,
+    indices_buffer: Option<A::Buffer>,
     texture: A::Texture,
     instances: [AccelerationStructureInstance; 3],
     instances_buffer: A::Buffer,
@@ -217,6 +224,18 @@ struct Example<A: hal::Api> {
 
 impl<A: hal::Api> Example<A> {
     fn init(window: &winit::window::Window) -> Result<Self, Box<dyn std::error::Error>> {
+        let mut index_buffer = false;
+
+        for arg in std::env::args() {
+            if arg == "index_buffer" {
+                index_buffer = true;
+            }
+        }
+
+        if index_buffer {
+            log::info!("using index buffer")
+        }
+
         let instance_desc = hal::InstanceDescriptor {
             name: "example",
             flags: wgt::InstanceFlags::default(),
@@ -420,29 +439,34 @@ impl<A: hal::Api> Example<A> {
             vertices_buffer
         };
 
-        let indices_buffer = unsafe {
-            let indices_buffer = device
-                .create_buffer(&hal::BufferDescriptor {
-                    label: Some("indices buffer"),
-                    size: indices_size_in_bytes as u64,
-                    usage: hal::BufferUses::MAP_WRITE
-                        | hal::BufferUses::BOTTOM_LEVEL_ACCELERATION_STRUCTURE_INPUT,
-                    memory_flags: hal::MemoryFlags::TRANSIENT | hal::MemoryFlags::PREFER_COHERENT,
-                })
-                .unwrap();
+        let indices_buffer = if index_buffer {
+            unsafe {
+                let indices_buffer = device
+                    .create_buffer(&hal::BufferDescriptor {
+                        label: Some("indices buffer"),
+                        size: indices_size_in_bytes as u64,
+                        usage: hal::BufferUses::MAP_WRITE
+                            | hal::BufferUses::BOTTOM_LEVEL_ACCELERATION_STRUCTURE_INPUT,
+                        memory_flags: hal::MemoryFlags::TRANSIENT
+                            | hal::MemoryFlags::PREFER_COHERENT,
+                    })
+                    .unwrap();
 
-            let mapping = device
-                .map_buffer(&indices_buffer, 0..indices_size_in_bytes as u64)
-                .unwrap();
-            ptr::copy_nonoverlapping(
-                indices.as_ptr() as *const u8,
-                mapping.ptr.as_ptr(),
-                indices_size_in_bytes,
-            );
-            device.unmap_buffer(&indices_buffer);
-            assert!(mapping.is_coherent);
+                let mapping = device
+                    .map_buffer(&indices_buffer, 0..indices_size_in_bytes as u64)
+                    .unwrap();
+                ptr::copy_nonoverlapping(
+                    indices.as_ptr() as *const u8,
+                    mapping.ptr.as_ptr(),
+                    indices_size_in_bytes,
+                );
+                device.unmap_buffer(&indices_buffer);
+                assert!(mapping.is_coherent);
 
-            indices_buffer
+                Some((indices_buffer, indices.len()))
+            }
+        } else {
+            None
         };
 
         let blas_triangles = vec![hal::AccelerationStructureTriangles {
@@ -451,12 +475,15 @@ impl<A: hal::Api> Example<A> {
             vertex_format: wgt::VertexFormat::Float32x3,
             vertex_count: vertices.len() as u32,
             vertex_stride: 3 * 4,
-            indices: Some(hal::AccelerationStructureTriangleIndices {
-                buffer: Some(&indices_buffer),
-                format: wgt::IndexFormat::Uint32,
-                offset: 0,
-                count: indices.len() as u32,
+            indices: indices_buffer.as_ref().map(|(buf, len)| {
+                hal::AccelerationStructureTriangleIndices {
+                    buffer: Some(buf),
+                    format: wgt::IndexFormat::Uint32,
+                    offset: 0,
+                    count: *len as u32,
+                }
             }),
+
             transform: None,
             flags: hal::AccelerationStructureGeometryFlags::OPAQUE,
         }];
@@ -800,7 +827,7 @@ impl<A: hal::Api> Example<A> {
             tlas,
             scratch_buffer,
             time: 0.0,
-            indices_buffer,
+            indices_buffer: indices_buffer.map(|(buf, _)| buf),
             vertices_buffer,
             uniform_buffer,
             texture_view,
@@ -1026,7 +1053,9 @@ impl<A: hal::Api> Example<A> {
             self.device.destroy_bind_group(self.bind_group);
             self.device.destroy_buffer(self.scratch_buffer);
             self.device.destroy_buffer(self.instances_buffer);
-            self.device.destroy_buffer(self.indices_buffer);
+            if let Some(buffer) = self.indices_buffer {
+                self.device.destroy_buffer(buffer);
+            }
             self.device.destroy_buffer(self.vertices_buffer);
             self.device.destroy_buffer(self.uniform_buffer);
             self.device.destroy_acceleration_structure(self.tlas);

third_party/rust/wgpu-hal/src/auxil/renderdoc.rs

@@ -74,7 +74,8 @@ impl RenderDoc {
                 Err(e) => {
                     return RenderDoc::NotAvailable {
                         reason: format!(
-                            "Unable to get RENDERDOC_GetAPI from renderdoc library '{renderdoc_filename}': {e:?}"
+                            "Unable to get RENDERDOC_GetAPI from renderdoc library '{}': {e:?}",
+                            renderdoc_filename
                         ),
                     }
                 }
@@ -89,7 +90,8 @@ impl RenderDoc {
             },
             return_value => RenderDoc::NotAvailable {
                 reason: format!(
-                    "Unable to get API from renderdoc library '{renderdoc_filename}': {return_value}"
+                    "Unable to get API from renderdoc library '{}': {}",
+                    renderdoc_filename, return_value
                 ),
             },
         }

third_party/rust/wgpu-hal/src/gles/mod.rs

@@ -457,11 +457,24 @@ impl Texture {
         };
 
         log::error!(
-            "wgpu-hal heuristics assumed that the view dimension will be equal to `{got}` rather than `{view_dimension:?}`.\n{}\n{}\n{}\n{}",
-            "`D2` textures with `depth_or_array_layers == 1` are assumed to have view dimension `D2`",
-            "`D2` textures with `depth_or_array_layers > 1` are assumed to have view dimension `D2Array`",
-            "`D2` textures with `depth_or_array_layers == 6` are assumed to have view dimension `Cube`",
-            "`D2` textures with `depth_or_array_layers > 6 && depth_or_array_layers % 6 == 0` are assumed to have view dimension `CubeArray`",
+            concat!(
+                "wgpu-hal heuristics assumed that ",
+                "the view dimension will be equal to `{}` rather than `{:?}`.\n",
+                "`D2` textures with ",
+                "`depth_or_array_layers == 1` ",
+                "are assumed to have view dimension `D2`\n",
+                "`D2` textures with ",
+                "`depth_or_array_layers > 1` ",
+                "are assumed to have view dimension `D2Array`\n",
+                "`D2` textures with ",
+                "`depth_or_array_layers == 6` ",
+                "are assumed to have view dimension `Cube`\n",
+                "`D2` textures with ",
+                "`depth_or_array_layers > 6 && depth_or_array_layers % 6 == 0` ",
+                "are assumed to have view dimension `CubeArray`\n",
+            ),
+            got,
+            view_dimension,
         );
     }
 }

third_party/rust/wgpu-hal/src/gles/queue.rs

@@ -1603,19 +1603,13 @@ impl super::Queue {
                 ref uniform,
                 offset,
             } => {
-                // T must be POD
-                //
-                // This function is absolutely sketchy and we really should be using bytemuck.
-                unsafe fn get_data<T, const COUNT: usize>(data: &[u8], offset: u32) -> &[T; COUNT] {
+                fn get_data<T, const COUNT: usize>(data: &[u8], offset: u32) -> [T; COUNT]
+                where
+                    [T; COUNT]: bytemuck::AnyBitPattern,
+                {
                     let data_required = size_of::<T>() * COUNT;
-
                     let raw = &data[(offset as usize)..][..data_required];
-
-                    debug_assert_eq!(data_required, raw.len());
-
-                    let slice: &[T] = unsafe { slice::from_raw_parts(raw.as_ptr().cast(), COUNT) };
-
-                    slice.try_into().unwrap()
+                    bytemuck::pod_read_unaligned(raw)
                 }
 
                 let location = Some(&uniform.location);
@@ -1625,28 +1619,28 @@ impl super::Queue {
                     // --- Float 1-4 Component ---
                     //
                     naga::TypeInner::Scalar(naga::Scalar::F32) => {
-                        let data = unsafe { get_data::<f32, 1>(data_bytes, offset)[0] };
+                        let data = get_data::<f32, 1>(data_bytes, offset)[0];
                         unsafe { gl.uniform_1_f32(location, data) };
                     }
                     naga::TypeInner::Vector {
                         size: naga::VectorSize::Bi,
                         scalar: naga::Scalar::F32,
                     } => {
-                        let data = unsafe { get_data::<f32, 2>(data_bytes, offset) };
+                        let data = &get_data::<f32, 2>(data_bytes, offset);
                         unsafe { gl.uniform_2_f32_slice(location, data) };
                     }
                     naga::TypeInner::Vector {
                         size: naga::VectorSize::Tri,
                         scalar: naga::Scalar::F32,
                     } => {
-                        let data = unsafe { get_data::<f32, 3>(data_bytes, offset) };
+                        let data = &get_data::<f32, 3>(data_bytes, offset);
                         unsafe { gl.uniform_3_f32_slice(location, data) };
                     }
                     naga::TypeInner::Vector {
                         size: naga::VectorSize::Quad,
                         scalar: naga::Scalar::F32,
                     } => {
-                        let data = unsafe { get_data::<f32, 4>(data_bytes, offset) };
+                        let data = &get_data::<f32, 4>(data_bytes, offset);
                         unsafe { gl.uniform_4_f32_slice(location, data) };
                     }
 
@@ -1654,28 +1648,28 @@ impl super::Queue {
                     // --- Int 1-4 Component ---
                     //
                     naga::TypeInner::Scalar(naga::Scalar::I32) => {
-                        let data = unsafe { get_data::<i32, 1>(data_bytes, offset)[0] };
+                        let data = get_data::<i32, 1>(data_bytes, offset)[0];
                         unsafe { gl.uniform_1_i32(location, data) };
                     }
                     naga::TypeInner::Vector {
                         size: naga::VectorSize::Bi,
                         scalar: naga::Scalar::I32,
                     } => {
-                        let data = unsafe { get_data::<i32, 2>(data_bytes, offset) };
+                        let data = &get_data::<i32, 2>(data_bytes, offset);
                         unsafe { gl.uniform_2_i32_slice(location, data) };
                     }
                     naga::TypeInner::Vector {
                         size: naga::VectorSize::Tri,
                         scalar: naga::Scalar::I32,
                     } => {
-                        let data = unsafe { get_data::<i32, 3>(data_bytes, offset) };
+                        let data = &get_data::<i32, 3>(data_bytes, offset);
                         unsafe { gl.uniform_3_i32_slice(location, data) };
                     }
                     naga::TypeInner::Vector {
                         size: naga::VectorSize::Quad,
                         scalar: naga::Scalar::I32,
                     } => {
-                        let data = unsafe { get_data::<i32, 4>(data_bytes, offset) };
+                        let data = &get_data::<i32, 4>(data_bytes, offset);
                         unsafe { gl.uniform_4_i32_slice(location, data) };
                     }
 
@@ -1683,28 +1677,28 @@ impl super::Queue {
                     // --- Uint 1-4 Component ---
                     //
                     naga::TypeInner::Scalar(naga::Scalar::U32) => {
-                        let data = unsafe { get_data::<u32, 1>(data_bytes, offset)[0] };
+                        let data = get_data::<u32, 1>(data_bytes, offset)[0];
                         unsafe { gl.uniform_1_u32(location, data) };
                     }
                     naga::TypeInner::Vector {
                         size: naga::VectorSize::Bi,
                         scalar: naga::Scalar::U32,
                     } => {
-                        let data = unsafe { get_data::<u32, 2>(data_bytes, offset) };
+                        let data = &get_data::<u32, 2>(data_bytes, offset);
                         unsafe { gl.uniform_2_u32_slice(location, data) };
                     }
                     naga::TypeInner::Vector {
                         size: naga::VectorSize::Tri,
                         scalar: naga::Scalar::U32,
                     } => {
-                        let data = unsafe { get_data::<u32, 3>(data_bytes, offset) };
+                        let data = &get_data::<u32, 3>(data_bytes, offset);
                         unsafe { gl.uniform_3_u32_slice(location, data) };
                     }
                     naga::TypeInner::Vector {
                         size: naga::VectorSize::Quad,
                         scalar: naga::Scalar::U32,
                     } => {
-                        let data = unsafe { get_data::<u32, 4>(data_bytes, offset) };
+                        let data = &get_data::<u32, 4>(data_bytes, offset);
                         unsafe { gl.uniform_4_u32_slice(location, data) };
                     }
 
@@ -1716,7 +1710,7 @@ impl super::Queue {
                         rows: naga::VectorSize::Bi,
                         scalar: naga::Scalar::F32,
                     } => {
-                        let data = unsafe { get_data::<f32, 4>(data_bytes, offset) };
+                        let data = &get_data::<f32, 4>(data_bytes, offset);
                         unsafe { gl.uniform_matrix_2_f32_slice(location, false, data) };
                     }
                     naga::TypeInner::Matrix {
@@ -1725,7 +1719,7 @@ impl super::Queue {
                         scalar: naga::Scalar::F32,
                     } => {
                         // repack 2 vec3s into 6 values.
-                        let unpacked_data = unsafe { get_data::<f32, 8>(data_bytes, offset) };
+                        let unpacked_data = &get_data::<f32, 8>(data_bytes, offset);
                         #[rustfmt::skip]
                         let packed_data = [
                             unpacked_data[0], unpacked_data[1], unpacked_data[2],
@@ -1738,7 +1732,7 @@ impl super::Queue {
                         rows: naga::VectorSize::Quad,
                         scalar: naga::Scalar::F32,
                     } => {
-                        let data = unsafe { get_data::<f32, 8>(data_bytes, offset) };
+                        let data = &get_data::<f32, 8>(data_bytes, offset);
                         unsafe { gl.uniform_matrix_2x4_f32_slice(location, false, data) };
                     }
 
@@ -1750,7 +1744,7 @@ impl super::Queue {
                         rows: naga::VectorSize::Bi,
                         scalar: naga::Scalar::F32,
                     } => {
-                        let data = unsafe { get_data::<f32, 6>(data_bytes, offset) };
+                        let data = &get_data::<f32, 6>(data_bytes, offset);
                         unsafe { gl.uniform_matrix_3x2_f32_slice(location, false, data) };
                     }
                     naga::TypeInner::Matrix {
@@ -1759,7 +1753,7 @@ impl super::Queue {
                         scalar: naga::Scalar::F32,
                     } => {
                         // repack 3 vec3s into 9 values.
-                        let unpacked_data = unsafe { get_data::<f32, 12>(data_bytes, offset) };
+                        let unpacked_data = &get_data::<f32, 12>(data_bytes, offset);
                         #[rustfmt::skip]
                         let packed_data = [
                             unpacked_data[0], unpacked_data[1], unpacked_data[2],
@@ -1773,7 +1767,7 @@ impl super::Queue {
                         rows: naga::VectorSize::Quad,
                         scalar: naga::Scalar::F32,
                     } => {
-                        let data = unsafe { get_data::<f32, 12>(data_bytes, offset) };
+                        let data = &get_data::<f32, 12>(data_bytes, offset);
                         unsafe { gl.uniform_matrix_3x4_f32_slice(location, false, data) };
                     }
 
@@ -1785,7 +1779,7 @@ impl super::Queue {
                         rows: naga::VectorSize::Bi,
                         scalar: naga::Scalar::F32,
                     } => {
-                        let data = unsafe { get_data::<f32, 8>(data_bytes, offset) };
+                        let data = &get_data::<f32, 8>(data_bytes, offset);
                         unsafe { gl.uniform_matrix_4x2_f32_slice(location, false, data) };
                     }
                     naga::TypeInner::Matrix {
@@ -1794,7 +1788,7 @@ impl super::Queue {
                         scalar: naga::Scalar::F32,
                     } => {
                         // repack 4 vec3s into 12 values.
-                        let unpacked_data = unsafe { get_data::<f32, 16>(data_bytes, offset) };
+                        let unpacked_data = &get_data::<f32, 16>(data_bytes, offset);
                         #[rustfmt::skip]
                         let packed_data = [
                             unpacked_data[0], unpacked_data[1], unpacked_data[2],
@@ -1809,7 +1803,7 @@ impl super::Queue {
                         rows: naga::VectorSize::Quad,
                         scalar: naga::Scalar::F32,
                     } => {
-                        let data = unsafe { get_data::<f32, 16>(data_bytes, offset) };
+                        let data = &get_data::<f32, 16>(data_bytes, offset);
                         unsafe { gl.uniform_matrix_4_f32_slice(location, false, data) };
                     }
                     _ => panic!("Unsupported uniform datatype: {:?}!", uniform.ty),

third_party/rust/wgpu-hal/src/gles/web.rs

@@ -64,9 +64,10 @@ impl Instance {
                 // “not supported” could include “insufficient GPU resources” or “the GPU process
                 // previously crashed”. So, we must return it as an `Err` since it could occur
                 // for circumstances outside the application author's control.
-                return Err(crate::InstanceError::new(String::from(
-                    "canvas.getContext() returned null; webgl2 not available or canvas already in use"
-                )));
+                return Err(crate::InstanceError::new(String::from(concat!(
+                    "canvas.getContext() returned null; ",
+                    "webgl2 not available or canvas already in use"
+                ))));
             }
             Err(js_error) => {
                 // <https://html.spec.whatwg.org/multipage/canvas.html#dom-canvas-getcontext>

third_party/rust/wgpu-hal/src/metal/command.rs

@@ -490,7 +490,7 @@ impl crate::CommandEncoder for super::CommandEncoder {
             wgt::QueryType::Timestamp => {
                 encoder.resolve_counters(
                     set.counter_sample_buffer.as_ref().unwrap(),
-                    metal::NSRange::new(range.start as u64, range.end as u64),
+                    metal::NSRange::new(range.start as u64, (range.end - range.start) as u64),
                     &buffer.raw,
                     offset,
                 );

third_party/rust/wgpu-hal/src/vulkan/command.rs

@@ -503,6 +503,9 @@ impl crate::CommandEncoder for super::CommandEncoder {
                     for triangles in in_geometries {
                         let mut triangle_data =
                             vk::AccelerationStructureGeometryTrianglesDataKHR::default()
+                                // IndexType::NONE_KHR is not set by default (due to being provided by VK_KHR_acceleration_structure) but unless there is an
+                                // index buffer we need to have IndexType::NONE_KHR as our index type.
+                                .index_type(vk::IndexType::NONE_KHR)
                                 .vertex_data(vk::DeviceOrHostAddressConstKHR {
                                     device_address: get_device_address(triangles.vertex_buffer),
                                 })

third_party/rust/wgpu-hal/src/vulkan/device.rs

@@ -2538,8 +2538,7 @@ impl super::DeviceShared {
                         }
                         None => {
                             crate::hal_usage_error(format!(
-                                "no signals reached value {}",
-                                wait_value
+                                "no signals reached value {wait_value}"
                             ));
                         }
                     }

third_party/rust/wgpu-hal/src/vulkan/instance.rs

@@ -745,7 +745,12 @@ impl crate::Instance for super::Instance {
                     Ok(sdk_ver) => sdk_ver,
                     Err(err) => {
                         log::error!(
-                            "Couldn't parse Android's ro.build.version.sdk system property ({val}): {err}"
+                            concat!(
+                                "Couldn't parse Android's ",
+                                "ro.build.version.sdk system property ({}): {}",
+                            ),
+                            val,
+                            err,
                         );
                         0
                     }
@@ -931,7 +936,10 @@ impl crate::Instance for super::Instance {
                         if version < (21, 2) {
                             // See https://gitlab.freedesktop.org/mesa/mesa/-/issues/4688
                             log::warn!(
-                                "Disabling presentation on '{}' (id {:?}) due to NV Optimus and Intel Mesa < v21.2",
+                                concat!(
+                                    "Disabling presentation on '{}' (id {:?}) ",
+                                    "due to NV Optimus and Intel Mesa < v21.2"
+                                ),
                                 exposed.info.name,
                                 exposed.adapter.raw
                             );

third_party/rust/wgpu-hal/src/vulkan/mod.rs

@@ -418,7 +418,13 @@ impl Surface {
             swapchain.next_present_time = Some(present_timing);
         } else {
             // Ideally we'd use something like `device.required_features` here, but that's in `wgpu-core`, which we are a dependency of
-            panic!("Tried to set display timing properties without the corresponding feature ({features:?}) enabled.");
+            panic!(
+                concat!(
+                    "Tried to set display timing properties ",
+                    "without the corresponding feature ({:?}) enabled."
+                ),
+                features
+            );
         }
     }
 }