Skip to content
This repository was archived by the owner on Oct 21, 2026. It is now read-only.

Commit bc444c8

Browse files
ChristopherBiscardiChristopherBiscardi
authored
wgpu 29 update (#134)
wgpu 29 is close to releasing, so this is a preview of the naga-29 changes for naga-oil. Changes are currently relative to: updating from [this commit](gfx-rs/wgpu@0ab6edd) to [the two fixes below](gfx-rs/wgpu@e80dc8d) causes two test failures ``` test compose::test::test::test_atomics ... FAILED test compose::test::test::use_shared_global ... FAILED ``` - [coherent and volatile memory decorations](gfx-rs/wgpu#9168) - [InstanceDescriptor::default no longer exists](gfx-rs/wgpu#8782)
1 parent e349f54 commit bc444c8

10 files changed

Lines changed: 71 additions & 38 deletions

File tree

‎Cargo.toml‎

Lines changed: 3 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -1,6 +1,6 @@
11
[package]
22
name = "naga_oil"
3-
version = "0.21.0"
3+
version = "0.22.0"
44
edition = "2021"
55
license = "MIT OR Apache-2.0"
66
description = "a crate for combining and manipulating shaders using naga IR"
@@ -18,7 +18,7 @@ prune = []
1818
allow_deprecated = []
1919

2020
[dependencies]
21-
naga = { version = "28", features = ["wgsl-in", "wgsl-out", "termcolor"] }
21+
naga = { version = "29", features = ["wgsl-in", "wgsl-out", "termcolor"] }
2222
tracing = "0.1"
2323
regex = "1.8"
2424
thiserror = "2.0"
@@ -29,6 +29,6 @@ unicode-ident = "1"
2929
indexmap = "2"
3030

3131
[dev-dependencies]
32-
wgpu = { version = "28", features = ["naga-ir"] }
32+
wgpu = { version = "29", features = ["naga-ir"] }
3333
futures-lite = "2"
3434
tracing-subscriber = { version = "0.3", features = ["std", "fmt"] }

‎examples/pbr_compose_test.rs‎

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -139,7 +139,7 @@ fn test_compose_final_module(n: usize, composer: &mut Composer) {
139139

140140
// make shader module from string
141141
fn test_wgsl_string_compile(n: usize) {
142-
let instance = wgpu::Instance::new(&wgpu::InstanceDescriptor::default());
142+
let instance = wgpu::Instance::new(wgpu::InstanceDescriptor::new_without_display_handle());
143143
let adapter =
144144
futures_lite::future::block_on(instance.enumerate_adapters(wgpu::Backends::all()))
145145
.into_iter()
@@ -162,7 +162,7 @@ fn test_wgsl_string_compile(n: usize) {
162162

163163
// make shader module from composed naga
164164
fn test_composer_compile(n: usize, composer: &mut Composer) {
165-
let instance = wgpu::Instance::new(&wgpu::InstanceDescriptor::default());
165+
let instance = wgpu::Instance::new(wgpu::InstanceDescriptor::new_without_display_handle());
166166
let adapter =
167167
futures_lite::future::block_on(instance.enumerate_adapters(wgpu::Backends::all()))
168168
.into_iter()

‎src/compose/mod.rs‎

Lines changed: 2 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -537,6 +537,7 @@ impl Composer {
537537
workgroup_size_overrides: None,
538538
mesh_info: None,
539539
task_payload: None,
540+
incoming_ray_payload: None,
540541
};
541542

542543
naga_module.entry_points.push(ep);
@@ -1835,6 +1836,7 @@ impl Composer {
18351836
workgroup_size_overrides: ep.workgroup_size_overrides,
18361837
mesh_info: ep.mesh_info.clone(),
18371838
task_payload: ep.task_payload,
1839+
incoming_ray_payload: ep.incoming_ray_payload,
18381840
});
18391841
}
18401842
let mut naga_module = naga::Module {

‎src/compose/test.rs‎

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -1488,7 +1488,7 @@ mod test {
14881488
})
14891489
.unwrap();
14901490

1491-
let instance = wgpu::Instance::new(&wgpu::InstanceDescriptor::default());
1491+
let instance = wgpu::Instance::new(wgpu::InstanceDescriptor::new_without_display_handle());
14921492
let adapter =
14931493
futures_lite::future::block_on(instance.enumerate_adapters(wgpu::Backends::all()))
14941494
.into_iter()
@@ -1532,7 +1532,7 @@ mod test {
15321532
layout: Some(
15331533
&device.create_pipeline_layout(&wgpu::PipelineLayoutDescriptor {
15341534
label: None,
1535-
bind_group_layouts: &[&layout],
1535+
bind_group_layouts: &[Some(&layout)],
15361536
immediate_size: 0,
15371537
}),
15381538
),

‎src/compose/tests/expected/atomics.txt‎

Lines changed: 17 additions & 17 deletions
Original file line numberDiff line numberDiff line change
@@ -6,26 +6,26 @@ fn entry_pointX_naga_oil_mod_XORSXG5C7NVXWI5LMMUX() -> f32 {
66
atomicStore((&atomX_naga_oil_mod_XORSXG5C7NVXWI5LMMUX), 1u);
77
let _e3: u32 = atomicLoad((&atomX_naga_oil_mod_XORSXG5C7NVXWI5LMMUX));
88
y = _e3;
9-
let _e7: u32 = atomicAdd((&atomX_naga_oil_mod_XORSXG5C7NVXWI5LMMUX), 2u);
10-
let _e8: u32 = y;
11-
y = (_e8 + _e7);
12-
let _e12: u32 = atomicSub((&atomX_naga_oil_mod_XORSXG5C7NVXWI5LMMUX), 1u);
13-
let _e13: u32 = y;
14-
y = (_e13 + _e12);
15-
let _e17: u32 = atomicMax((&atomX_naga_oil_mod_XORSXG5C7NVXWI5LMMUX), 5u);
16-
let _e18: u32 = y;
17-
y = (_e18 + _e17);
18-
let _e22: u32 = atomicMin((&atomX_naga_oil_mod_XORSXG5C7NVXWI5LMMUX), 4u);
19-
let _e23: u32 = y;
20-
y = (_e23 + _e22);
9+
let _e5: u32 = y;
10+
let _e8: u32 = atomicAdd((&atomX_naga_oil_mod_XORSXG5C7NVXWI5LMMUX), 2u);
11+
y = (_e5 + _e8);
12+
let _e10: u32 = y;
13+
let _e13: u32 = atomicSub((&atomX_naga_oil_mod_XORSXG5C7NVXWI5LMMUX), 1u);
14+
y = (_e10 + _e13);
15+
let _e15: u32 = y;
16+
let _e18: u32 = atomicMax((&atomX_naga_oil_mod_XORSXG5C7NVXWI5LMMUX), 5u);
17+
y = (_e15 + _e18);
18+
let _e20: u32 = y;
19+
let _e23: u32 = atomicMin((&atomX_naga_oil_mod_XORSXG5C7NVXWI5LMMUX), 4u);
20+
y = (_e20 + _e23);
2121
let _e25: u32 = y;
22-
let _e27: u32 = atomicExchange((&atomX_naga_oil_mod_XORSXG5C7NVXWI5LMMUX), _e25);
23-
let _e28: u32 = y;
24-
y = (_e28 + _e27);
22+
let _e26: u32 = y;
23+
let _e28: u32 = atomicExchange((&atomX_naga_oil_mod_XORSXG5C7NVXWI5LMMUX), _e26);
24+
y = (_e25 + _e28);
2525
let _e33: _atomic_compare_exchange_result_Uint_4_ = atomicCompareExchangeWeak((&atomX_naga_oil_mod_XORSXG5C7NVXWI5LMMUX), 12u, 0u);
2626
if _e33.exchanged {
27-
let _e36: u32 = y;
28-
y = (_e36 + _e33.old_value);
27+
let _e35: u32 = y;
28+
y = (_e35 + _e33.old_value);
2929
}
3030
let _e38: u32 = y;
3131
return f32(_e38);

‎src/compose/tests/expected/err_validation_1.txt‎

Lines changed: 2 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -4,7 +4,8 @@ error: failed to build a valid final module: Function [1] 'func' is invalid
44
7 │ ╭ fn func() -> f32 {
55
8 │ │ return 1u;
66
│ │ ^^ naga::ir::Expression [0]
7-
│ ╰──────────────^ naga::ir::Function [1]
7+
9 │ │ }
8+
│ ╰─^ naga::ir::Function [1]
89
│
910
= The `return` expression Some([0]) does not match the declared return type Some([0])
1011

‎src/compose/tests/expected/err_validation_2.txt‎

Lines changed: 2 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -4,7 +4,8 @@ error: failed to build a valid final module: Function [0] 'valid_inc::func' is i
44
7 │ ╭ fn func() -> f32 {
55
8 │ │ return 1u;
66
│ │ ^^ naga::ir::Expression [0]
7-
│ ╰──────────────^ naga::ir::Function [0]
7+
9 │ │ }
8+
│ ╰─^ naga::ir::Function [0]
89
│
910
= The `return` expression Some([0]) does not match the declared return type Some([0])
1011

‎src/compose/tests/expected/use_shared_global.txt‎

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -1,8 +1,8 @@
11
var<private> aX_naga_oil_mod_XNVXWIX: f32 = 0f;
22

33
fn add() {
4-
let _e2: f32 = aX_naga_oil_mod_XNVXWIX;
5-
aX_naga_oil_mod_XNVXWIX = (_e2 + 1f);
4+
let _e1: f32 = aX_naga_oil_mod_XNVXWIX;
5+
aX_naga_oil_mod_XNVXWIX = (_e1 + 1f);
66
return;
77
}
88

‎src/derive.rs‎

Lines changed: 36 additions & 9 deletions
Original file line numberDiff line numberDiff line change
@@ -131,7 +131,6 @@ impl<'a> DerivedModule<'a> {
131131
| TypeInner::Atomic { .. }
132132
| TypeInner::AccelerationStructure { .. }
133133
| TypeInner::RayQuery { .. } => ty.inner.clone(),
134-
135134
TypeInner::Pointer { base, space } => TypeInner::Pointer {
136135
base: self.import_type(base),
137136
space: *space,
@@ -160,6 +159,17 @@ impl<'a> DerivedModule<'a> {
160159
base: self.import_type(base),
161160
size: *size,
162161
},
162+
TypeInner::CooperativeMatrix {
163+
columns,
164+
rows,
165+
scalar,
166+
role,
167+
} => TypeInner::CooperativeMatrix {
168+
columns: *columns,
169+
rows: *rows,
170+
scalar: *scalar,
171+
role: *role,
172+
},
163173
},
164174
};
165175
let span = self.shader.as_ref().unwrap().types.get_span(*h_type);
@@ -212,6 +222,7 @@ impl<'a> DerivedModule<'a> {
212222
binding: gv.binding,
213223
ty: self.import_type(&gv.ty),
214224
init: gv.init.map(|c| self.import_global_expression(c)),
225+
memory_decorations: gv.memory_decorations,
215226
};
216227

217228
let span = self
@@ -319,7 +330,6 @@ impl<'a> DerivedModule<'a> {
319330
.iter()
320331
.map(|stmt| {
321332
match stmt {
322-
// remap function calls
323333
Statement::Call {
324334
function,
325335
arguments,
@@ -329,8 +339,6 @@ impl<'a> DerivedModule<'a> {
329339
arguments: arguments.iter().map(|expr| map_expr!(expr)).collect(),
330340
result: result.as_ref().map(|result| map_expr!(result)),
331341
},
332-
333-
// recursively
334342
Statement::Block(b) => Statement::Block(map_block!(b)),
335343
Statement::If {
336344
condition,
@@ -361,8 +369,6 @@ impl<'a> DerivedModule<'a> {
361369
continuing: map_block!(continuing),
362370
break_if: map_expr_opt!(break_if),
363371
},
364-
365-
// map expressions
366372
Statement::Emit(exprs) => {
367373
// iterate once to add expressions that should NOT be part of the emit statement
368374
for expr in exprs.clone() {
@@ -513,12 +519,18 @@ impl<'a> DerivedModule<'a> {
513519
value: map_expr!(value),
514520
}
515521
}
516-
// else just copy
517522
Statement::Break
518523
| Statement::Continue
519524
| Statement::Kill
520525
| Statement::MemoryBarrier(_)
521526
| Statement::ControlBarrier(_) => stmt.clone(),
527+
Statement::RayPipelineFunction(ray_pipeline_function) => {
528+
Statement::RayPipelineFunction(*ray_pipeline_function)
529+
}
530+
Statement::CooperativeStore { target, data } => Statement::CooperativeStore {
531+
target: map_expr!(target),
532+
data: *data,
533+
},
522534
}
523535
})
524536
.collect();
@@ -722,12 +734,10 @@ impl<'a> DerivedModule<'a> {
722734
convert: *convert,
723735
},
724736
Expression::ArrayLength(expr) => Expression::ArrayLength(map_expr!(expr)),
725-
726737
Expression::LocalVariable(_) | Expression::FunctionArgument(_) => {
727738
is_external = true;
728739
expr.clone()
729740
}
730-
731741
Expression::AtomicResult { ty, comparison } => Expression::AtomicResult {
732742
ty: self.import_type(ty),
733743
comparison: *comparison,
@@ -758,6 +768,22 @@ impl<'a> DerivedModule<'a> {
758768
Expression::SubgroupOperationResult { ty } => Expression::SubgroupOperationResult {
759769
ty: self.import_type(ty),
760770
},
771+
Expression::CooperativeLoad {
772+
columns,
773+
rows,
774+
role,
775+
data,
776+
} => Expression::CooperativeLoad {
777+
columns: *columns,
778+
rows: *rows,
779+
role: *role,
780+
data: *data,
781+
},
782+
Expression::CooperativeMultiplyAdd { a, b, c } => Expression::CooperativeMultiplyAdd {
783+
a: map_expr!(a),
784+
b: map_expr!(b),
785+
c: map_expr!(c),
786+
},
761787
};
762788

763789
if !non_emitting_only || is_external {
@@ -910,6 +936,7 @@ impl<'a> DerivedModule<'a> {
910936
workgroup_size_overrides: ep.workgroup_size_overrides,
911937
mesh_info: ep.mesh_info.clone(),
912938
task_payload: ep.task_payload,
939+
incoming_ray_payload: ep.incoming_ray_payload,
913940
})
914941
.collect();
915942

‎src/redirect.rs‎

Lines changed: 3 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -73,7 +73,9 @@ impl Redirector {
7373
| Statement::SubgroupBallot { .. }
7474
| Statement::SubgroupGather { .. }
7575
| Statement::SubgroupCollectiveOperation { .. }
76-
| Statement::ImageAtomic { .. } => (),
76+
| Statement::ImageAtomic { .. }
77+
| Statement::RayPipelineFunction(..)
78+
| Statement::CooperativeStore { .. } => (),
7779
}
7880
}
7981
}

0 commit comments

Comments
 (0)