[INFO] cloning repository https://github.com/Pauan/rust-wgpu-compute [INFO] running `Command { std: "git" "-c" "credential.helper=" "-c" "credential.helper=/workspace/cargo-home/bin/git-credential-null" "clone" "--bare" "https://github.com/Pauan/rust-wgpu-compute" "/workspace/cache/git-repos/https%3A%2F%2Fgithub.com%2FPauan%2Frust-wgpu-compute", kill_on_drop: false }` [INFO] [stderr] Cloning into bare repository '/workspace/cache/git-repos/https%3A%2F%2Fgithub.com%2FPauan%2Frust-wgpu-compute'... [INFO] running `Command { std: "git" "rev-parse" "HEAD", kill_on_drop: false }` [INFO] [stdout] 5c3cbbb50dc7a3f4bf642b0331aaa8f8566065ec [INFO] checking Pauan/rust-wgpu-compute against try#a611f2a14e38407ec6717a86a01424ee6fc80762 for pr-154992-1 [INFO] running `Command { std: "git" "clone" "/workspace/cache/git-repos/https%3A%2F%2Fgithub.com%2FPauan%2Frust-wgpu-compute" "/workspace/builds/worker-6-tc2/source", kill_on_drop: false }` [INFO] [stderr] Cloning into '/workspace/builds/worker-6-tc2/source'... [INFO] [stderr] done. [INFO] removed /workspace/builds/worker-6-tc2/source/.cargo/config.toml [INFO] started tweaking git repo https://github.com/Pauan/rust-wgpu-compute [INFO] finished tweaking git repo https://github.com/Pauan/rust-wgpu-compute [INFO] tweaked toml for git repo https://github.com/Pauan/rust-wgpu-compute written to /workspace/builds/worker-6-tc2/source/Cargo.toml [INFO] validating manifest of git repo https://github.com/Pauan/rust-wgpu-compute on toolchain a611f2a14e38407ec6717a86a01424ee6fc80762 [INFO] running `Command { std: CARGO_HOME="/workspace/cargo-home" RUSTUP_HOME="/workspace/rustup-home" "/workspace/cargo-home/bin/cargo" "+a611f2a14e38407ec6717a86a01424ee6fc80762" "metadata" "--manifest-path" "Cargo.toml" "--no-deps", kill_on_drop: false }` [INFO] crate git repo https://github.com/Pauan/rust-wgpu-compute already has a lockfile, it will not be regenerated [INFO] running `Command { std: CARGO_HOME="/workspace/cargo-home" RUSTUP_HOME="/workspace/rustup-home" "/workspace/cargo-home/bin/cargo" "+a611f2a14e38407ec6717a86a01424ee6fc80762" "fetch" "--manifest-path" "Cargo.toml", kill_on_drop: false }` [INFO] running `Command { std: "docker" "create" "-v" "/var/lib/crater-agent-workspace/builds/worker-6-tc2/target:/opt/rustwide/target:rw,Z" "-v" "/var/lib/crater-agent-workspace/builds/worker-6-tc2/source:/opt/rustwide/workdir:ro,Z" "-v" "/var/lib/crater-agent-workspace/cargo-home:/opt/rustwide/cargo-home:ro,Z" "-v" "/var/lib/crater-agent-workspace/rustup-home:/opt/rustwide/rustup-home:ro,Z" "-e" "SOURCE_DIR=/opt/rustwide/workdir" "-e" "CARGO_TARGET_DIR=/opt/rustwide/target" "-e" "CARGO_HOME=/opt/rustwide/cargo-home" "-e" "RUSTUP_HOME=/opt/rustwide/rustup-home" "-w" "/opt/rustwide/workdir" "-m" "1610612736" "--user" "0:0" "--network" "none" "ghcr.io/rust-lang/crates-build-env/linux@sha256:d429b63d4308055ea97f60fb1d3dfca48854a00942f1bd2ad806beaf015945ec" "/opt/rustwide/cargo-home/bin/cargo" "+a611f2a14e38407ec6717a86a01424ee6fc80762" "metadata" "--no-deps" "--format-version=1", kill_on_drop: false }` [INFO] [stdout] c581d84e3c3aedb09c196b282d069b529511fe9340197de96f9d02a2e4d59f3c [INFO] running `Command { std: "docker" "start" "-a" "c581d84e3c3aedb09c196b282d069b529511fe9340197de96f9d02a2e4d59f3c", kill_on_drop: false }` [INFO] running `Command { std: "docker" "inspect" "c581d84e3c3aedb09c196b282d069b529511fe9340197de96f9d02a2e4d59f3c", kill_on_drop: false }` [INFO] running `Command { std: "docker" "rm" "-f" "c581d84e3c3aedb09c196b282d069b529511fe9340197de96f9d02a2e4d59f3c", kill_on_drop: false }` [INFO] [stdout] c581d84e3c3aedb09c196b282d069b529511fe9340197de96f9d02a2e4d59f3c [INFO] running `Command { std: "docker" "create" "-v" "/var/lib/crater-agent-workspace/builds/worker-6-tc2/target:/opt/rustwide/target:rw,Z" "-v" "/var/lib/crater-agent-workspace/builds/worker-6-tc2/source:/opt/rustwide/workdir:ro,Z" "-v" "/var/lib/crater-agent-workspace/cargo-home:/opt/rustwide/cargo-home:ro,Z" "-v" "/var/lib/crater-agent-workspace/rustup-home:/opt/rustwide/rustup-home:ro,Z" "-e" "SOURCE_DIR=/opt/rustwide/workdir" "-e" "CARGO_TARGET_DIR=/opt/rustwide/target" "-e" "CARGO_INCREMENTAL=0" "-e" "RUST_BACKTRACE=full" "-e" "RUSTFLAGS=--cap-lints=forbid" "-e" "RUSTDOCFLAGS=--cap-lints=forbid" "-e" "CARGO_HOME=/opt/rustwide/cargo-home" "-e" "RUSTUP_HOME=/opt/rustwide/rustup-home" "-w" "/opt/rustwide/workdir" "-m" "1610612736" "--user" "0:0" "--network" "none" "ghcr.io/rust-lang/crates-build-env/linux@sha256:d429b63d4308055ea97f60fb1d3dfca48854a00942f1bd2ad806beaf015945ec" "/opt/rustwide/cargo-home/bin/cargo" "+a611f2a14e38407ec6717a86a01424ee6fc80762" "check" "--frozen" "--all" "--all-targets" "--message-format=json", kill_on_drop: false }` [INFO] [stdout] b2818a9bfe306eae81688112ac7b120744a54b7286b4f1e19f7ed8120244d75c [INFO] running `Command { std: "docker" "start" "-a" "b2818a9bfe306eae81688112ac7b120744a54b7286b4f1e19f7ed8120244d75c", kill_on_drop: false }` [INFO] [stderr] Compiling proc-macro2 v1.0.106 [INFO] [stderr] Compiling quote v1.0.44 [INFO] [stderr] Compiling unicode-ident v1.0.22 [INFO] [stderr] Compiling libm v0.2.16 [INFO] [stderr] Compiling num-traits v0.2.19 [INFO] [stderr] Compiling zerocopy v0.8.38 [INFO] [stderr] Compiling thiserror v2.0.18 [INFO] [stderr] Checking once_cell v1.21.3 [INFO] [stderr] Checking log v0.4.29 [INFO] [stderr] Compiling wasm-bindgen-shared v0.2.108 [INFO] [stderr] Compiling libc v0.2.180 [INFO] [stderr] Checking codespan-reporting v0.12.0 [INFO] [stderr] Checking foldhash v0.2.0 [INFO] [stderr] Checking libloading v0.8.9 [INFO] [stderr] Compiling naga v28.0.0 [INFO] [stderr] Checking hashbrown v0.16.1 [INFO] [stderr] Compiling naga v26.0.0 [INFO] [stderr] Compiling bumpalo v3.19.1 [INFO] [stderr] Checking ash v0.38.0+1.3.281 [INFO] [stderr] Compiling wasm-bindgen v0.2.108 [INFO] [stderr] Checking gpu-descriptor-types v0.2.0 [INFO] [stderr] Checking spirv v0.3.0+sdk-1.3.268.0 [INFO] [stderr] Compiling wgpu-hal v28.0.0 [INFO] [stderr] Checking presser v0.3.1 [INFO] [stderr] Checking futures-core v0.3.31 [INFO] [stderr] Checking futures-sink v0.3.31 [INFO] [stderr] Checking futures-channel v0.3.31 [INFO] [stderr] Checking gpu-descriptor v0.3.2 [INFO] [stderr] Compiling syn v2.0.114 [INFO] [stderr] Checking indexmap v2.13.0 [INFO] [stderr] Checking profiling v1.0.17 [INFO] [stderr] Compiling unicode-width v0.2.2 [INFO] [stderr] Checking futures-io v0.3.31 [INFO] [stderr] Checking ordered-float v5.0.0 [INFO] [stderr] Compiling wgpu-core v28.0.0 [INFO] [stderr] Compiling bit-vec v0.8.0 [INFO] [stderr] Checking parking_lot_core v0.9.12 [INFO] [stderr] Checking khronos-egl v6.0.0 [INFO] [stderr] Compiling bit-set v0.8.0 [INFO] [stderr] Checking parking_lot v0.12.5 [INFO] [stderr] Compiling hashbrown v0.15.5 [INFO] [stderr] Compiling wgpu v28.0.0 [INFO] [stderr] Compiling arrayvec v0.7.6 [INFO] [stderr] Compiling bitflags v2.10.0 [INFO] [stderr] Compiling termcolor v1.4.1 [INFO] [stderr] Compiling hexf-parse v0.2.1 [INFO] [stderr] Compiling rustc-hash v1.1.0 [INFO] [stderr] Compiling rayon-core v1.13.0 [INFO] [stderr] Compiling codespan-reporting v0.13.1 [INFO] [stderr] Checking rayon v1.11.0 [INFO] [stderr] Compiling wasm-bindgen-macro-support v0.2.108 [INFO] [stderr] Compiling zerocopy-derive v0.8.38 [INFO] [stderr] Compiling thiserror-impl v2.0.18 [INFO] [stderr] Compiling bytemuck_derive v1.10.2 [INFO] [stderr] Compiling futures-macro v0.3.31 [INFO] [stderr] Checking futures-util v0.3.31 [INFO] [stderr] Checking bytemuck v1.25.0 [INFO] [stderr] Checking wgpu-types v28.0.0 [INFO] [stderr] Compiling wasm-bindgen-macro v0.2.108 [INFO] [stderr] Checking js-sys v0.3.85 [INFO] [stderr] Checking console_error_panic_hook v0.1.7 [INFO] [stderr] Checking futures-executor v0.3.31 [INFO] [stderr] Checking futures v0.3.31 [INFO] [stderr] Checking half v2.7.1 [INFO] [stderr] Checking web-sys v0.3.85 [INFO] [stderr] Checking wasm-bindgen-futures v0.4.58 [INFO] [stderr] Checking console_log v1.0.0 [INFO] [stderr] Checking gpu-allocator v0.28.0 [INFO] [stderr] Checking wgpu-compute-macro v0.1.0 (/opt/rustwide/workdir/crates/wgpu-compute-macro) [INFO] [stdout] warning: unreachable statement [INFO] [stdout] --> crates/wgpu-compute-macro/src/generate.rs:403:17 [INFO] [stdout] | [INFO] [stdout] 402 | todo!(); [INFO] [stdout] | ------- any code following this expression is unreachable [INFO] [stdout] 403 | unreachable!(); [INFO] [stdout] | ^^^^^^^^^^^^^^ unreachable statement [INFO] [stdout] | [INFO] [stdout] = note: `#[warn(unreachable_code)]` (part of `#[warn(unused)]`) on by default [INFO] [stdout] [INFO] [stdout] [INFO] [stdout] warning: unreachable statement [INFO] [stdout] --> crates/wgpu-compute-macro/src/generate.rs:403:17 [INFO] [stdout] | [INFO] [stdout] 402 | todo!(); [INFO] [stdout] | ------- any code following this expression is unreachable [INFO] [stdout] 403 | unreachable!(); [INFO] [stdout] | ^^^^^^^^^^^^^^ unreachable statement [INFO] [stdout] | [INFO] [stdout] = note: `#[warn(unreachable_code)]` (part of `#[warn(unused)]`) on by default [INFO] [stdout] [INFO] [stdout] [INFO] [stdout] warning: unused variable: `handle` [INFO] [stdout] --> crates/wgpu-compute-macro/src/generate.rs:392:35 [INFO] [stdout] | [INFO] [stdout] 392 | Expression::ZeroValue(handle) => { [INFO] [stdout] | ^^^^^^ help: if this is intentional, prefix it with an underscore: `_handle` [INFO] [stdout] | [INFO] [stdout] = note: `#[warn(unused_variables)]` (part of `#[warn(unused)]`) on by default [INFO] [stdout] [INFO] [stdout] [INFO] [stdout] warning: unused variable: `size` [INFO] [stdout] --> crates/wgpu-compute-macro/src/generate.rs:396:33 [INFO] [stdout] | [INFO] [stdout] 396 | Expression::Splat { size, value } => { [INFO] [stdout] | ^^^^ help: try ignoring the field: `size: _` [INFO] [stdout] [INFO] [stdout] [INFO] [stdout] warning: unused variable: `value` [INFO] [stdout] --> crates/wgpu-compute-macro/src/generate.rs:396:39 [INFO] [stdout] | [INFO] [stdout] 396 | Expression::Splat { size, value } => { [INFO] [stdout] | ^^^^^ help: try ignoring the field: `value: _` [INFO] [stdout] [INFO] [stdout] [INFO] [stdout] warning: unused variable: `handle` [INFO] [stdout] --> crates/wgpu-compute-macro/src/generate.rs:392:35 [INFO] [stdout] | [INFO] [stdout] 392 | Expression::ZeroValue(handle) => { [INFO] [stdout] | ^^^^^^ help: if this is intentional, prefix it with an underscore: `_handle` [INFO] [stdout] | [INFO] [stdout] = note: `#[warn(unused_variables)]` (part of `#[warn(unused)]`) on by default [INFO] [stdout] [INFO] [stdout] [INFO] [stdout] warning: unused variable: `size` [INFO] [stdout] --> crates/wgpu-compute-macro/src/generate.rs:396:33 [INFO] [stdout] | [INFO] [stdout] 396 | Expression::Splat { size, value } => { [INFO] [stdout] | ^^^^ help: try ignoring the field: `size: _` [INFO] [stdout] [INFO] [stdout] [INFO] [stdout] warning: unused variable: `value` [INFO] [stdout] --> crates/wgpu-compute-macro/src/generate.rs:396:39 [INFO] [stdout] | [INFO] [stdout] 396 | Expression::Splat { size, value } => { [INFO] [stdout] | ^^^^^ help: try ignoring the field: `value: _` [INFO] [stdout] [INFO] [stdout] [INFO] [stdout] warning: field `filename` is never read [INFO] [stdout] --> crates/wgpu-compute-macro/src/lib.rs:32:5 [INFO] [stdout] | [INFO] [stdout] 30 | struct File { [INFO] [stdout] | ---- field in this struct [INFO] [stdout] 31 | span: Span, [INFO] [stdout] 32 | filename: String, [INFO] [stdout] | ^^^^^^^^ [INFO] [stdout] | [INFO] [stdout] = note: `#[warn(dead_code)]` (part of `#[warn(unused)]`) on by default [INFO] [stdout] [INFO] [stdout] [INFO] [stdout] warning: field `filename` is never read [INFO] [stdout] --> crates/wgpu-compute-macro/src/lib.rs:32:5 [INFO] [stdout] | [INFO] [stdout] 30 | struct File { [INFO] [stdout] | ---- field in this struct [INFO] [stdout] 31 | span: Span, [INFO] [stdout] 32 | filename: String, [INFO] [stdout] | ^^^^^^^^ [INFO] [stdout] | [INFO] [stdout] = note: `#[warn(dead_code)]` (part of `#[warn(unused)]`) on by default [INFO] [stdout] [INFO] [stdout] [INFO] [stdout] warning: unreachable statement [INFO] [stdout] --> crates/wgpu-compute-macro/src/generate.rs:403:17 [INFO] [stdout] | [INFO] [stdout] 402 | todo!(); [INFO] [stdout] | ------- any code following this expression is unreachable [INFO] [stdout] 403 | unreachable!(); [INFO] [stdout] | ^^^^^^^^^^^^^^ unreachable statement [INFO] [stdout] | [INFO] [stdout] = note: `#[warn(unreachable_code)]` (part of `#[warn(unused)]`) on by default [INFO] [stdout] [INFO] [stdout] [INFO] [stdout] warning: unused variable: `handle` [INFO] [stdout] --> crates/wgpu-compute-macro/src/generate.rs:392:35 [INFO] [stdout] | [INFO] [stdout] 392 | Expression::ZeroValue(handle) => { [INFO] [stdout] | ^^^^^^ help: if this is intentional, prefix it with an underscore: `_handle` [INFO] [stdout] | [INFO] [stdout] = note: `#[warn(unused_variables)]` (part of `#[warn(unused)]`) on by default [INFO] [stdout] [INFO] [stdout] [INFO] [stdout] warning: unused variable: `size` [INFO] [stdout] --> crates/wgpu-compute-macro/src/generate.rs:396:33 [INFO] [stdout] | [INFO] [stdout] 396 | Expression::Splat { size, value } => { [INFO] [stdout] | ^^^^ help: try ignoring the field: `size: _` [INFO] [stdout] [INFO] [stdout] [INFO] [stdout] warning: unused variable: `value` [INFO] [stdout] --> crates/wgpu-compute-macro/src/generate.rs:396:39 [INFO] [stdout] | [INFO] [stdout] 396 | Expression::Splat { size, value } => { [INFO] [stdout] | ^^^^^ help: try ignoring the field: `value: _` [INFO] [stdout] [INFO] [stdout] [INFO] [stdout] warning: field `filename` is never read [INFO] [stdout] --> crates/wgpu-compute-macro/src/lib.rs:32:5 [INFO] [stdout] | [INFO] [stdout] 30 | struct File { [INFO] [stdout] | ---- field in this struct [INFO] [stdout] 31 | span: Span, [INFO] [stdout] 32 | filename: String, [INFO] [stdout] | ^^^^^^^^ [INFO] [stdout] | [INFO] [stdout] = note: `#[warn(dead_code)]` (part of `#[warn(unused)]`) on by default [INFO] [stdout] [INFO] [stdout] [INFO] [stderr] Checking wgpu-core-deps-windows-linux-android v28.0.0 [INFO] [stderr] Checking wgpu-compute v0.1.0 (/opt/rustwide/workdir/crates/wgpu-compute) [INFO] [stderr] Checking gpu v0.1.0 (/opt/rustwide/workdir) [INFO] [stdout] #[derive(:: std :: fmt :: Debug, :: std :: clone :: Clone, :: std :: marker :: [INFO] [stdout] Copy, :: bytemuck :: NoUninit, :: bytemuck :: AnyBitPattern,)] #[repr(C)] pub [INFO] [stdout] struct Input { pub value : f32, } impl Input [INFO] [stdout] { pub fn new(value : f32) -> Self { Self { value } } } [INFO] [stdout] #[derive(:: std :: fmt :: Debug, :: std :: clone :: Clone, :: std :: marker :: [INFO] [stdout] Copy, :: bytemuck :: NoUninit, :: bytemuck :: AnyBitPattern,)] #[repr(C)] pub [INFO] [stdout] struct Output { pub value : f32, } impl Output [INFO] [stdout] { pub fn new(value : f32) -> Self { Self { value } } } pub const FOO : Input = [INFO] [stdout] Input { value : 3f32, }; pub const BAR : [f32; 2usize] = [5f32, 1f32]; pub [INFO] [stdout] const QUX : [[f32; 3usize]; 2usize] = [INFO] [stdout] [[1f32, 2f32, 3f32], [4f32, 5f32, 6f32]]; pub struct Bindings [INFO] [stdout] { [INFO] [stdout] pub input : :: std :: vec :: Vec < Input > , pub output : :: std :: vec :: [INFO] [stdout] Vec < Output > , [INFO] [stdout] } pub struct CpuBuffers [INFO] [stdout] { [INFO] [stdout] pub input : :: std :: sync :: Mutex < :: std :: vec :: Vec < Input > > , [INFO] [stdout] pub output : :: std :: sync :: Mutex < :: std :: vec :: Vec < Output > > , [INFO] [stdout] } pub struct GpuBuffers [INFO] [stdout] { [INFO] [stdout] pub input : :: wgpu_compute :: Input < :: std :: vec :: Vec < Input > > , [INFO] [stdout] pub output : :: wgpu_compute :: Input < :: std :: vec :: Vec < Output > > [INFO] [stdout] , [INFO] [stdout] } impl GpuBuffers [INFO] [stdout] { [INFO] [stdout] :: std :: thread_local! [INFO] [stdout] { [INFO] [stdout] static GPU_LAYOUT : :: std :: cell :: OnceCell < :: wgpu_compute :: [INFO] [stdout] __internal :: GpuLayout > = :: std :: cell :: OnceCell :: new(); [INFO] [stdout] } fn gpu_layout < 'a > [INFO] [stdout] (gpu : & :: wgpu_compute :: Gpu, layout : & 'a :: std :: cell :: OnceCell [INFO] [stdout] < :: wgpu_compute :: __internal :: GpuLayout > ,) -> & 'a :: wgpu_compute [INFO] [stdout] :: __internal :: GpuLayout [INFO] [stdout] { [INFO] [stdout] layout.get_or_init(|| [INFO] [stdout] { [INFO] [stdout] :: wgpu_compute :: __internal :: GpuLayout :: [INFO] [stdout] new(gpu, [INFO] [stdout] "struct Input {\n value: f32,\n}\n\nstruct Output {\n value: f32,\n}\n\nconst FOO: Input = Input(3f);\nconst BAR: vec2 = vec2(5f, 1f);\nconst QUX: mat2x3 = mat2x3(vec3(1f, 2f, 3f), vec3(4f, 5f, 6f));\n\n@group(0) @binding(0) \nvar input: array;\n@group(0) @binding(1) \nvar output: array;\n\nfn double_1(value: f32) -> f32 {\n return (value * 2f);\n}\n\n@compute @workgroup_size(64, 1, 1) \nfn double(@builtin(global_invocation_id) global_id: vec3) {\n let index: u32 = global_id.x;\n if (index >= arrayLength((&input))) {\n return;\n }\n let _e11: f32 = input[index].value;\n let _e12: f32 = double_1(_e11);\n output[index].value = _e12;\n return;\n}\n", [INFO] [stdout] & [INFO] [stdout] [& [INFO] [stdout] [:: wgpu :: BindGroupLayoutEntry [INFO] [stdout] { [INFO] [stdout] binding : 0u32, visibility : :: wgpu :: ShaderStages :: [INFO] [stdout] COMPUTE, ty : :: wgpu :: BindingType :: Buffer [INFO] [stdout] { [INFO] [stdout] ty : :: wgpu :: BufferBindingType :: Storage [INFO] [stdout] { read_only : true }, has_dynamic_offset : false, [INFO] [stdout] min_binding_size : :: std :: option :: Option :: [INFO] [stdout] Some(:: std :: num :: NonZeroU64 :: [INFO] [stdout] new(:: std :: mem :: size_of :: < Input > () as [INFO] [stdout] u64).unwrap()), [INFO] [stdout] }, count : :: std :: option :: Option :: None, [INFO] [stdout] }, :: wgpu :: BindGroupLayoutEntry [INFO] [stdout] { [INFO] [stdout] binding : 1u32, visibility : :: wgpu :: ShaderStages :: [INFO] [stdout] COMPUTE, ty : :: wgpu :: BindingType :: Buffer [INFO] [stdout] { [INFO] [stdout] ty : :: wgpu :: BufferBindingType :: Storage [INFO] [stdout] { read_only : false }, has_dynamic_offset : false, [INFO] [stdout] min_binding_size : :: std :: option :: Option :: [INFO] [stdout] Some(:: std :: num :: NonZeroU64 :: [INFO] [stdout] new(:: std :: mem :: size_of :: < Output > () as [INFO] [stdout] u64).unwrap()), [INFO] [stdout] }, count : :: std :: option :: Option :: None, [INFO] [stdout] },]],) [INFO] [stdout] }) [INFO] [stdout] } fn bind_group < A, F > (& self, f : F) -> A where F : [INFO] [stdout] FnOnce(& [& [:: wgpu :: BindGroupEntry]]) -> A [INFO] [stdout] { [INFO] [stdout] f(& [INFO] [stdout] [& [INFO] [stdout] [:: wgpu :: BindGroupEntry [INFO] [stdout] { [INFO] [stdout] binding : 0u32, resource : :: wgpu_compute :: Gpu :: [INFO] [stdout] bind_group(& self.input), [INFO] [stdout] }, :: wgpu :: BindGroupEntry [INFO] [stdout] { [INFO] [stdout] binding : 1u32, resource : :: wgpu_compute :: Gpu :: [INFO] [stdout] bind_group(& self.output), [INFO] [stdout] },]]) [INFO] [stdout] } [INFO] [stdout] } impl :: wgpu_compute :: IntoBuffers for Bindings [INFO] [stdout] { [INFO] [stdout] type Cpu = CpuBuffers; type Gpu = GpuBuffers; fn into_cpu_buffers(self) -> [INFO] [stdout] Self :: Cpu [INFO] [stdout] { [INFO] [stdout] CpuBuffers [INFO] [stdout] { [INFO] [stdout] input : :: std :: sync :: Mutex :: new(self.input), output : :: [INFO] [stdout] std :: sync :: Mutex :: new(self.output), [INFO] [stdout] } [INFO] [stdout] } fn into_gpu_buffers(self, gpu : & :: wgpu_compute :: Gpu) -> Self :: Gpu [INFO] [stdout] { [INFO] [stdout] GpuBuffers [INFO] [stdout] { [INFO] [stdout] input : gpu.input_vec(self.input.as_slice()), output : [INFO] [stdout] gpu.input_vec(self.output.as_slice()), [INFO] [stdout] } [INFO] [stdout] } [INFO] [stdout] } struct CpuState < 'a > [INFO] [stdout] { [INFO] [stdout] state : & 'a :: wgpu_compute :: __internal :: StateCpu < CpuBuffers > , [INFO] [stdout] global_id : [u32; 3], [INFO] [stdout] } impl < 'a > CpuState < 'a > [INFO] [stdout] { [INFO] [stdout] #[inline] fn write_input(& self) -> :: std :: sync :: MutexGuard < '_, :: [INFO] [stdout] std :: vec :: Vec < Input > > { self.state.buffers.input.lock().unwrap() } [INFO] [stdout] #[inline] fn write_output(& self) -> :: std :: sync :: MutexGuard < '_, :: [INFO] [stdout] std :: vec :: Vec < Output > > [INFO] [stdout] { self.state.buffers.output.lock().unwrap() } [INFO] [stdout] } fn double__cpu_impl < 'a > (__state__ : & CpuState < 'a > , value : f32) -> [INFO] [stdout] f32 { return value * 2f32; } fn double_cpu_impl < 'a > [INFO] [stdout] (__state__ : & CpuState < 'a >) [INFO] [stdout] { [INFO] [stdout] let index = __state__.global_id [0usize]; if index >= [INFO] [stdout] (__state__.write_input().len() as u32) { return; } else {} let v12 = [INFO] [stdout] double__cpu_impl(__state__, __state__.write_input() [INFO] [stdout] [index as usize].value); __state__.write_output() [index as usize].value = [INFO] [stdout] v12; return; [INFO] [stdout] } pub fn [INFO] [stdout] double_cpu(state : :: wgpu_compute :: StateCpu < CpuBuffers > , threads : [INFO] [stdout] u32,) [INFO] [stdout] { [INFO] [stdout] use :: rayon :: iter :: { ParallelIterator, IntoParallelIterator }; let [INFO] [stdout] state : :: wgpu_compute :: __internal :: StateCpu < CpuBuffers > = :: std [INFO] [stdout] :: convert :: Into :: into(state); [INFO] [stdout] (0u32 .. [INFO] [stdout] threads).into_par_iter().for_each(| index | [INFO] [stdout] { [INFO] [stdout] double_cpu_impl(& CpuState [INFO] [stdout] { state : & state, global_id : [index, 0, 0], }) [INFO] [stdout] }); [INFO] [stdout] } pub fn [INFO] [stdout] double_gpu(state : :: wgpu_compute :: StateGpu < GpuBuffers > , threads : [INFO] [stdout] u32,) -> impl :: std :: future :: Future < Output = () > + use < > [INFO] [stdout] { [INFO] [stdout] let state : :: wgpu_compute :: __internal :: StateGpu < GpuBuffers > = :: [INFO] [stdout] std :: convert :: Into :: into(state); :: std :: thread_local! [INFO] [stdout] { [INFO] [stdout] static GPU_FN : :: std :: cell :: OnceCell < :: wgpu_compute :: [INFO] [stdout] __internal :: GpuFn > = :: std :: cell :: OnceCell :: new(); [INFO] [stdout] } GpuBuffers :: [INFO] [stdout] GPU_LAYOUT.with(| gpu_layout | [INFO] [stdout] { [INFO] [stdout] let gpu_layout = GpuBuffers :: gpu_layout(& state.gpu, gpu_layout); [INFO] [stdout] GPU_FN.with(| gpu_fn | [INFO] [stdout] { [INFO] [stdout] let gpu_fn = [INFO] [stdout] gpu_fn.get_or_init(|| [INFO] [stdout] { [INFO] [stdout] :: wgpu_compute :: __internal :: GpuFn :: [INFO] [stdout] new(& state.gpu, gpu_layout, [64u32, 1u32, 1u32], "double",) [INFO] [stdout] }); let mut encoder = [INFO] [stdout] state.buffers.bind_group(| bind_group | [INFO] [stdout] { [INFO] [stdout] :: wgpu_compute :: __internal :: [INFO] [stdout] command_encoder(& state.gpu, gpu_layout, gpu_fn, threads, [INFO] [stdout] bind_group,) [INFO] [stdout] }); for (input, output) in state.copy_buffers [INFO] [stdout] { [INFO] [stdout] :: wgpu_compute :: __internal :: [INFO] [stdout] copy_input_to_output(& mut encoder, & input, & output); [INFO] [stdout] } :: wgpu_compute :: __internal :: wait(& state.gpu, encoder) [INFO] [stdout] }) [INFO] [stdout] }) [INFO] [stdout] } pub fn double(state : :: wgpu_compute :: State < Bindings > , threads : u32) [INFO] [stdout] -> impl :: std :: future :: Future < Output = () > + use < > [INFO] [stdout] { [INFO] [stdout] async move [INFO] [stdout] { [INFO] [stdout] match state [INFO] [stdout] { [INFO] [stdout] :: wgpu_compute :: State :: Gpu(state) => [INFO] [stdout] { double_gpu(state, threads).await }, :: wgpu_compute :: State :: [INFO] [stdout] Cpu(state) => { double_cpu(state, threads) }, [INFO] [stdout] } [INFO] [stdout] } [INFO] [stdout] } [INFO] [stdout] #[derive(:: std :: fmt :: Debug, :: std :: clone :: Clone, :: std :: marker :: [INFO] [stdout] Copy, :: bytemuck :: NoUninit, :: bytemuck :: AnyBitPattern,)] #[repr(C)] pub [INFO] [stdout] struct Input { pub value : f32, } impl Input [INFO] [stdout] { pub fn new(value : f32) -> Self { Self { value } } } [INFO] [stdout] #[derive(:: std :: fmt :: Debug, :: std :: clone :: Clone, :: std :: marker :: [INFO] [stdout] Copy, :: bytemuck :: NoUninit, :: bytemuck :: AnyBitPattern,)] #[repr(C)] pub [INFO] [stdout] struct Output { pub value : f32, } impl Output [INFO] [stdout] { pub fn new(value : f32) -> Self { Self { value } } } pub const FOO : Input = [INFO] [stdout] Input { value : 3f32, }; pub const BAR : [f32; 2usize] = [5f32, 1f32]; pub [INFO] [stdout] const QUX : [[f32; 3usize]; 2usize] = [INFO] [stdout] [[1f32, 2f32, 3f32], [4f32, 5f32, 6f32]]; pub struct Bindings [INFO] [stdout] { [INFO] [stdout] pub input : :: std :: vec :: Vec < Input > , pub output : :: std :: vec :: [INFO] [stdout] Vec < Output > , [INFO] [stdout] } pub struct CpuBuffers [INFO] [stdout] { [INFO] [stdout] pub input : :: std :: sync :: Mutex < :: std :: vec :: Vec < Input > > , [INFO] [stdout] pub output : :: std :: sync :: Mutex < :: std :: vec :: Vec < Output > > , [INFO] [stdout] } pub struct GpuBuffers [INFO] [stdout] { [INFO] [stdout] pub input : :: wgpu_compute :: Input < :: std :: vec :: Vec < Input > > , [INFO] [stdout] pub output : :: wgpu_compute :: Input < :: std :: vec :: Vec < Output > > [INFO] [stdout] , [INFO] [stdout] } impl GpuBuffers [INFO] [stdout] { [INFO] [stdout] :: std :: thread_local! [INFO] [stdout] { [INFO] [stdout] static GPU_LAYOUT : :: std :: cell :: OnceCell < :: wgpu_compute :: [INFO] [stdout] __internal :: GpuLayout > = :: std :: cell :: OnceCell :: new(); [INFO] [stdout] } fn gpu_layout < 'a > [INFO] [stdout] (gpu : & :: wgpu_compute :: Gpu, layout : & 'a :: std :: cell :: OnceCell [INFO] [stdout] < :: wgpu_compute :: __internal :: GpuLayout > ,) -> & 'a :: wgpu_compute [INFO] [stdout] :: __internal :: GpuLayout [INFO] [stdout] { [INFO] [stdout] layout.get_or_init(|| [INFO] [stdout] { [INFO] [stdout] :: wgpu_compute :: __internal :: GpuLayout :: [INFO] [stdout] new(gpu, [INFO] [stdout] "struct Input {\n value: f32,\n}\n\nstruct Output {\n value: f32,\n}\n\nconst FOO: Input = Input(3f);\nconst BAR: vec2 = vec2(5f, 1f);\nconst QUX: mat2x3 = mat2x3(vec3(1f, 2f, 3f), vec3(4f, 5f, 6f));\n\n@group(0) @binding(0) \nvar input: array;\n@group(0) @binding(1) \nvar output: array;\n\nfn double_1(value: f32) -> f32 {\n return (value * 2f);\n}\n\n@compute @workgroup_size(64, 1, 1) \nfn double(@builtin(global_invocation_id) global_id: vec3) {\n let index: u32 = global_id.x;\n if (index >= arrayLength((&input))) {\n return;\n }\n let _e11: f32 = input[index].value;\n let _e12: f32 = double_1(_e11);\n output[index].value = _e12;\n return;\n}\n", [INFO] [stdout] & [INFO] [stdout] [& [INFO] [stdout] [:: wgpu :: BindGroupLayoutEntry [INFO] [stdout] { [INFO] [stdout] binding : 0u32, visibility : :: wgpu :: ShaderStages :: [INFO] [stdout] COMPUTE, ty : :: wgpu :: BindingType :: Buffer [INFO] [stdout] { [INFO] [stdout] ty : :: wgpu :: BufferBindingType :: Storage [INFO] [stdout] { read_only : true }, has_dynamic_offset : false, [INFO] [stdout] min_binding_size : :: std :: option :: Option :: [INFO] [stdout] Some(:: std :: num :: NonZeroU64 :: [INFO] [stdout] new(:: std :: mem :: size_of :: < Input > () as [INFO] [stdout] u64).unwrap()), [INFO] [stdout] }, count : :: std :: option :: Option :: None, [INFO] [stdout] }, :: wgpu :: BindGroupLayoutEntry [INFO] [stdout] { [INFO] [stdout] binding : 1u32, visibility : :: wgpu :: ShaderStages :: [INFO] [stdout] COMPUTE, ty : :: wgpu :: BindingType :: Buffer [INFO] [stdout] { [INFO] [stdout] ty : :: wgpu :: BufferBindingType :: Storage [INFO] [stdout] { read_only : false }, has_dynamic_offset : false, [INFO] [stdout] min_binding_size : :: std :: option :: Option :: [INFO] [stdout] Some(:: std :: num :: NonZeroU64 :: [INFO] [stdout] new(:: std :: mem :: size_of :: < Output > () as [INFO] [stdout] u64).unwrap()), [INFO] [stdout] }, count : :: std :: option :: Option :: None, [INFO] [stdout] },]],) [INFO] [stdout] }) [INFO] [stdout] } fn bind_group < A, F > (& self, f : F) -> A where F : [INFO] [stdout] FnOnce(& [& [:: wgpu :: BindGroupEntry]]) -> A [INFO] [stdout] { [INFO] [stdout] f(& [INFO] [stdout] [& [INFO] [stdout] [:: wgpu :: BindGroupEntry [INFO] [stdout] { [INFO] [stdout] binding : 0u32, resource : :: wgpu_compute :: Gpu :: [INFO] [stdout] bind_group(& self.input), [INFO] [stdout] }, :: wgpu :: BindGroupEntry [INFO] [stdout] { [INFO] [stdout] binding : 1u32, resource : :: wgpu_compute :: Gpu :: [INFO] [stdout] bind_group(& self.output), [INFO] [stdout] },]]) [INFO] [stdout] } [INFO] [stdout] } impl :: wgpu_compute :: IntoBuffers for Bindings [INFO] [stdout] { [INFO] [stdout] type Cpu = CpuBuffers; type Gpu = GpuBuffers; fn into_cpu_buffers(self) -> [INFO] [stdout] Self :: Cpu [INFO] [stdout] { [INFO] [stdout] CpuBuffers [INFO] [stdout] { [INFO] [stdout] input : :: std :: sync :: Mutex :: new(self.input), output : :: [INFO] [stdout] std :: sync :: Mutex :: new(self.output), [INFO] [stdout] } [INFO] [stdout] } fn into_gpu_buffers(self, gpu : & :: wgpu_compute :: Gpu) -> Self :: Gpu [INFO] [stdout] { [INFO] [stdout] GpuBuffers [INFO] [stdout] { [INFO] [stdout] input : gpu.input_vec(self.input.as_slice()), output : [INFO] [stdout] gpu.input_vec(self.output.as_slice()), [INFO] [stdout] } [INFO] [stdout] } [INFO] [stdout] } struct CpuState < 'a > [INFO] [stdout] { [INFO] [stdout] state : & 'a :: wgpu_compute :: __internal :: StateCpu < CpuBuffers > , [INFO] [stdout] global_id : [u32; 3], [INFO] [stdout] } impl < 'a > CpuState < 'a > [INFO] [stdout] { [INFO] [stdout] #[inline] fn write_input(& self) -> :: std :: sync :: MutexGuard < '_, :: [INFO] [stdout] std :: vec :: Vec < Input > > { self.state.buffers.input.lock().unwrap() } [INFO] [stdout] #[inline] fn write_output(& self) -> :: std :: sync :: MutexGuard < '_, :: [INFO] [stdout] std :: vec :: Vec < Output > > [INFO] [stdout] { self.state.buffers.output.lock().unwrap() } [INFO] [stdout] } fn double__cpu_impl < 'a > (__state__ : & CpuState < 'a > , value : f32) -> [INFO] [stdout] f32 { return value * 2f32; } fn double_cpu_impl < 'a > [INFO] [stdout] (__state__ : & CpuState < 'a >) [INFO] [stdout] { [INFO] [stdout] let index = __state__.global_id [0usize]; if index >= [INFO] [stdout] (__state__.write_input().len() as u32) { return; } else {} let v12 = [INFO] [stdout] double__cpu_impl(__state__, __state__.write_input() [INFO] [stdout] [index as usize].value); __state__.write_output() [index as usize].value = [INFO] [stdout] v12; return; [INFO] [stdout] } pub fn [INFO] [stdout] double_cpu(state : :: wgpu_compute :: StateCpu < CpuBuffers > , threads : [INFO] [stdout] u32,) [INFO] [stdout] { [INFO] [stdout] use :: rayon :: iter :: { ParallelIterator, IntoParallelIterator }; let [INFO] [stdout] state : :: wgpu_compute :: __internal :: StateCpu < CpuBuffers > = :: std [INFO] [stdout] :: convert :: Into :: into(state); [INFO] [stdout] (0u32 .. [INFO] [stdout] threads).into_par_iter().for_each(| index | [INFO] [stdout] { [INFO] [stdout] double_cpu_impl(& CpuState [INFO] [stdout] { state : & state, global_id : [index, 0, 0], }) [INFO] [stdout] }); [INFO] [stdout] } pub fn [INFO] [stdout] double_gpu(state : :: wgpu_compute :: StateGpu < GpuBuffers > , threads : [INFO] [stdout] u32,) -> impl :: std :: future :: Future < Output = () > + use < > [INFO] [stdout] { [INFO] [stdout] let state : :: wgpu_compute :: __internal :: StateGpu < GpuBuffers > = :: [INFO] [stdout] std :: convert :: Into :: into(state); :: std :: thread_local! [INFO] [stdout] { [INFO] [stdout] static GPU_FN : :: std :: cell :: OnceCell < :: wgpu_compute :: [INFO] [stdout] __internal :: GpuFn > = :: std :: cell :: OnceCell :: new(); [INFO] [stdout] } GpuBuffers :: [INFO] [stdout] GPU_LAYOUT.with(| gpu_layout | [INFO] [stdout] { [INFO] [stdout] let gpu_layout = GpuBuffers :: gpu_layout(& state.gpu, gpu_layout); [INFO] [stdout] GPU_FN.with(| gpu_fn | [INFO] [stdout] { [INFO] [stdout] let gpu_fn = [INFO] [stdout] gpu_fn.get_or_init(|| [INFO] [stdout] { [INFO] [stdout] :: wgpu_compute :: __internal :: GpuFn :: [INFO] [stdout] new(& state.gpu, gpu_layout, [64u32, 1u32, 1u32], "double",) [INFO] [stdout] }); let mut encoder = [INFO] [stdout] state.buffers.bind_group(| bind_group | [INFO] [stdout] { [INFO] [stdout] :: wgpu_compute :: __internal :: [INFO] [stdout] command_encoder(& state.gpu, gpu_layout, gpu_fn, threads, [INFO] [stdout] bind_group,) [INFO] [stdout] }); for (input, output) in state.copy_buffers [INFO] [stdout] { [INFO] [stdout] :: wgpu_compute :: __internal :: [INFO] [stdout] copy_input_to_output(& mut encoder, & input, & output); [INFO] [stdout] } :: wgpu_compute :: __internal :: wait(& state.gpu, encoder) [INFO] [stdout] }) [INFO] [stdout] }) [INFO] [stdout] } pub fn double(state : :: wgpu_compute :: State < Bindings > , threads : u32) [INFO] [stdout] -> impl :: std :: future :: Future < Output = () > + use < > [INFO] [stdout] { [INFO] [stdout] async move [INFO] [stdout] { [INFO] [stdout] match state [INFO] [stdout] { [INFO] [stdout] :: wgpu_compute :: State :: Gpu(state) => [INFO] [stdout] { double_gpu(state, threads).await }, :: wgpu_compute :: State :: [INFO] [stdout] Cpu(state) => { double_cpu(state, threads) }, [INFO] [stdout] } [INFO] [stdout] } [INFO] [stdout] } [INFO] [stdout] warning: variable does not need to be mutable [INFO] [stdout] --> src/lib.rs:15:9 [INFO] [stdout] | [INFO] [stdout] 15 | let mut state = wgpu_compute::State::new(gpu::Bindings { [INFO] [stdout] | ----^^^^^ [INFO] [stdout] | | [INFO] [stdout] | help: remove this `mut` [INFO] [stdout] | [INFO] [stdout] = note: `#[warn(unused_mut)]` (part of `#[warn(unused)]`) on by default [INFO] [stdout] [INFO] [stdout] [INFO] [stdout] warning: variable does not need to be mutable [INFO] [stdout] --> src/lib.rs:15:9 [INFO] [stdout] | [INFO] [stdout] 15 | let mut state = wgpu_compute::State::new(gpu::Bindings { [INFO] [stdout] | ----^^^^^ [INFO] [stdout] | | [INFO] [stdout] | help: remove this `mut` [INFO] [stdout] | [INFO] [stdout] = note: `#[warn(unused_mut)]` (part of `#[warn(unused)]`) on by default [INFO] [stdout] [INFO] [stdout] [INFO] [stderr] Finished `dev` profile [unoptimized + debuginfo] target(s) in 1m 48s [INFO] running `Command { std: "docker" "inspect" "b2818a9bfe306eae81688112ac7b120744a54b7286b4f1e19f7ed8120244d75c", kill_on_drop: false }` [INFO] running `Command { std: "docker" "rm" "-f" "b2818a9bfe306eae81688112ac7b120744a54b7286b4f1e19f7ed8120244d75c", kill_on_drop: false }` [INFO] [stdout] b2818a9bfe306eae81688112ac7b120744a54b7286b4f1e19f7ed8120244d75c