Skip to content

Commit 6187aca

Browse files
committed
Add isolated wgpu validation gate
1 parent ebbccde commit 6187aca

14 files changed

Lines changed: 502 additions & 0 deletions

.github/workflows/check.yml

Lines changed: 7 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -41,6 +41,7 @@ jobs:
4141
~/.cargo/git
4242
~/.cargo/registry
4343
tools/naga_oil_oracle/target
44+
tools/wgpu_validation/.mooncakes/Milky2018/wgpu_mbt/src/c/target
4445
key: ${{ runner.os }}-cargo-${{ hashFiles('tools/naga_oil_oracle/Cargo.lock') }}
4546
restore-keys: |
4647
${{ runner.os }}-cargo-
@@ -54,6 +55,7 @@ jobs:
5455
- name: Moon package update
5556
run: |
5657
moon update
58+
moon -C tools/wgpu_validation update
5759
5860
- name: Moon check
5961
run: |
@@ -88,6 +90,11 @@ jobs:
8890
ulimit -s 8176
8991
bash tools/check_wgsl_validation.sh
9092
93+
- name: wgpu runtime validation
94+
run: |
95+
ulimit -s 8176
96+
bash tools/check_wgpu_validation.sh
97+
9198
- name: Moon WGSL byte parity
9299
run: |
93100
ulimit -s 8176

docs/moon_wgsl-issue-tracker.md

Lines changed: 1 addition & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -13,6 +13,7 @@ Last updated: 2026-05-07
1313

1414
| ID | Source | Problem | Status | Notes |
1515
| --- | --- | --- | --- | --- |
16+
| `WGSL-217` | Validation architecture review | The validation gate still depended on Naga as the deepest executable oracle. Naga validates WGSL syntax and many IR semantics, but it does not fully exercise wgpu pipeline layout compatibility, so runtime binding-access mismatches such as read-only storage layouts versus read-write shader globals could regress without a native WebGPU gate. Directly adding `wgpu_mbt` to the root module would contaminate the published library with a native-only dependency. | `DONE` | Added an isolated `tools/wgpu_validation` MoonBit subproject with its own `moon.mod.json` and native-only `Milky2018/wgpu_mbt` dependency. Added `tools/check_wgpu_validation.sh`, which builds the subproject binary, validates shader-module creation for generated Bevy PBR and MGStudio forward shaders, IR-roundtrips a storage-access fixture, validates it against an explicit read-only/writable storage pipeline layout, and includes a negative control that proves wgpu rejects read-write shader access against a read-only layout. CI now updates the isolated subproject and runs the wgpu gate without adding `wgpu_mbt` to the root module dependency graph. Verified locally on 2026-05-07. |
1617
| `WGSL-216` | Architecture risk review | The transform reference rewrite plan still exposed symbol bindings as `(rel_path, original_name, from_name, to_name)` arguments at the public boundary, so callers could construct an identity-shaped binding without actually carrying the canonical IR identity object. The validation improvements from `WGSL-215` also had no single guardrail that would fail if future changes reintroduced CTS allowlists, cached alias phases, unscoped rewrite APIs, or dropped real Bevy/MGStudio IR validation. | `DONE` | Changed `WgslReferenceRewritePlan::add_symbol_binding` to require `WgslIrSymbolIdentity` directly and changed transform rename rules to keep optional identities instead of separate provenance strings. Added `tools/check_architecture_guardrails.sh`, which rejects the old official CTS allowlist path, cached qualified alias binding phases, public unscoped string rewrite APIs, reference rewrite bindings with split provenance strings, and missing full Bevy PBR / MGStudio IR validation gates. Unsupported expression lowering now reports explicit AST node categories instead of a catch-all fallback. Verified locally on 2026-05-07. |
1718
| `WGSL-215` | Architecture risk review | The validation surface still allowed three structural blind spots: the official CTS IR gate depended on a handwritten allowlist, real Bevy/MGStudio forward shaders were validated only as composed source rather than as `compose -> IR -> emit` output, and composer still exposed a public unscoped string reference-rewrite API plus a cached alias binding phase distinct from ordinary identity bindings. | `DONE` | Removed `testdata/gpuweb_cts_ir_allowlist.txt`; the official corpus gate now requires every extracted static-valid case to parse, lower to IR, emit WGSL, and parse again, with only Naga validation skipped for oracle-blocked subgroup-enable cases. Added Bevy PBR functions, full Bevy PBR forward, and MGStudio mesh3d forward compose outputs to the WGSL validation gate, each with IR roundtrip plus Naga validation. Removed public `WgslReferenceRewritePlan::add`, deleted `CachedQualifiedAliasBinding`, and made alias binding dedup preserve same-identity final-name correction while rejecting cross-identity alias collisions. Verified locally on 2026-05-07. |
1819
| `WGSL-214` | Parser / IR structural coverage | Real Bevy PBR IR roundtrip exposed two core WGSL lowering gaps: the statement parser could attach `continuing` following an `if` block to the `if` statement instead of keeping it as the loop continuing clause, and IR lvalue lowering did not treat `ptr<function, T>` function arguments as addressable values for WGSL's automatic pointer-member access/writeback. | `DONE` | Restricted statement continuation so `else` only continues `if` statements and `continuing` only continues `loop` statements. Changed loop lowering to split a `continuing` clause out of the loop body's statement list into the IR loop continuing block. Added pointer function-argument lvalue lowering and parser regressions for nested `if` conditions plus loop continuing ownership. Verified with full Bevy PBR/MGStudio compose-to-IR validation on 2026-05-07. |

moon.mod.json

Lines changed: 3 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -15,5 +15,8 @@
1515
"wgsl",
1616
"shader"
1717
],
18+
"exclude": [
19+
"tools/wgpu_validation"
20+
],
1821
"description": "WGSL preprocess and composer utilities aligned with naga_oil."
1922
}

tools/check_architecture_guardrails.sh

Lines changed: 10 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -45,6 +45,16 @@ if ! rg -n 'roundtrip_and_validate_wgsl "\$tmpdir/mgstudio_mesh3d_forward\.wgsl"
4545
fail "WGSL validation gate must IR-roundtrip MGStudio mesh3d forward"
4646
fi
4747

48+
if ! rg -n 'bash tools/check_wgpu_validation\.sh' .github/workflows/check.yml >/dev/null; then
49+
fail "CI must run native wgpu runtime validation"
50+
fi
51+
52+
if ! rg -n 'compute-storage-read' tools/check_wgpu_validation.sh tools/wgpu_validation \
53+
--glob '!tools/wgpu_validation/_build/**' \
54+
--glob '!tools/wgpu_validation/.mooncakes/**' >/dev/null; then
55+
fail "wgpu validation must include explicit read-only storage layout coverage"
56+
fi
57+
4858
if ! rg -n 'moon run tools/ir_roundtrip -- --input "\$case_file" --output "\$emitted"' tools/check_official_wgsl_corpus.sh >/dev/null; then
4959
fail "official WGSL CTS gate must lower every extracted case through IR"
5060
fi

tools/check_wgpu_validation.sh

Lines changed: 93 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,93 @@
1+
#!/usr/bin/env bash
2+
set -euo pipefail
3+
4+
repo_root="$(cd "$(dirname "${BASH_SOURCE[0]}")/.." && pwd)"
5+
cd "$repo_root"
6+
7+
tmpdir="$(mktemp -d)"
8+
negative_output="$tmpdir/negative.out"
9+
cleanup() {
10+
rm -rf "$tmpdir"
11+
}
12+
trap cleanup EXIT
13+
14+
moon -C tools/wgpu_validation build --target native >/dev/null
15+
validator="$repo_root/tools/wgpu_validation/_build/native/debug/build/cmd/main/main.exe"
16+
17+
wgpu_validate() {
18+
"$validator" "$@"
19+
}
20+
21+
emit_case() {
22+
local case_name="$1"
23+
local output="$2"
24+
moon run tools/wgsl_validation_cases -- "$case_name" > "$output"
25+
}
26+
27+
echo "== wgpu validation: shader module smoke =="
28+
cat > "$tmpdir/shader_module_smoke.wgsl" <<'WGSL'
29+
@fragment
30+
fn fs_main() -> @location(0) vec4<f32> {
31+
return vec4<f32>(0.0, 0.0, 0.0, 1.0);
32+
}
33+
WGSL
34+
wgpu_validate --input "$tmpdir/shader_module_smoke.wgsl" --mode shader-module
35+
36+
echo "== wgpu validation: storage access runtime layout =="
37+
cat > "$tmpdir/storage_access_source.wgsl" <<'WGSL'
38+
@group(0) @binding(0)
39+
var<storage> values: array<u32>;
40+
41+
@group(0) @binding(1)
42+
var<storage, read_write> output: array<u32>;
43+
44+
@compute @workgroup_size(1)
45+
fn main(@builtin(global_invocation_id) id: vec3<u32>) {
46+
if (id.x >= arrayLength(&values)) {
47+
return;
48+
}
49+
output[id.x] = values[id.x];
50+
}
51+
WGSL
52+
moon run tools/ir_roundtrip -- \
53+
--input "$tmpdir/storage_access_source.wgsl" \
54+
--output "$tmpdir/storage_access_ir.wgsl" >/dev/null
55+
wgpu_validate \
56+
--input "$tmpdir/storage_access_ir.wgsl" \
57+
--mode compute-storage-read \
58+
--compute-entry main
59+
60+
echo "== wgpu validation: storage access negative control =="
61+
cat > "$tmpdir/storage_access_bad.wgsl" <<'WGSL'
62+
@group(0) @binding(0)
63+
var<storage, read_write> values: array<u32>;
64+
65+
@group(0) @binding(1)
66+
var<storage, read_write> output: array<u32>;
67+
68+
@compute @workgroup_size(1)
69+
fn main(@builtin(global_invocation_id) id: vec3<u32>) {
70+
if (id.x >= arrayLength(&values)) {
71+
return;
72+
}
73+
output[id.x] = values[id.x];
74+
}
75+
WGSL
76+
if wgpu_validate \
77+
--input "$tmpdir/storage_access_bad.wgsl" \
78+
--mode compute-storage-read \
79+
--compute-entry main >"$negative_output" 2>&1; then
80+
cat "$negative_output" >&2
81+
echo "expected wgpu validation to reject read_write shader binding against read-only layout" >&2
82+
exit 1
83+
fi
84+
85+
echo "== wgpu validation: Bevy PBR forward shader module =="
86+
emit_case bevy-pbr-forward "$tmpdir/bevy_pbr_forward.wgsl"
87+
wgpu_validate --input "$tmpdir/bevy_pbr_forward.wgsl" --mode shader-module
88+
89+
echo "== wgpu validation: MGStudio mesh3d forward shader module =="
90+
emit_case mgstudio-mesh3d-forward "$tmpdir/mgstudio_mesh3d_forward.wgsl"
91+
wgpu_validate --input "$tmpdir/mgstudio_mesh3d_forward.wgsl" --mode shader-module
92+
93+
echo "wgpu validation gate passed"

tools/wgpu_validation/.gitignore

Lines changed: 4 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,4 @@
1+
.DS_Store
2+
_build/
3+
.mooncakes/
4+
.moonagent/
Lines changed: 34 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,34 @@
1+
# moon_wgsl wgpu validation harness
2+
3+
This is an isolated native-only MoonBit subproject for validating generated
4+
WGSL with `Milky2018/wgpu_mbt`.
5+
6+
It intentionally lives outside the root `Milky2018/moon_wgsl` module so the
7+
published WGSL library remains backend-agnostic and can still be checked for
8+
all MoonBit targets.
9+
10+
## Usage
11+
12+
```bash
13+
moon -C tools/wgpu_validation build --target native
14+
tools/wgpu_validation/_build/native/debug/build/cmd/main/main.exe \
15+
--input path/to/shader.wgsl \
16+
--mode shader-module
17+
```
18+
19+
Modes:
20+
21+
- `shader-module`: create a wgpu shader module and inspect compilation info.
22+
- `compute`: create a compute pipeline with the selected entry point.
23+
- `compute-storage-read`: create a compute pipeline with an explicit layout
24+
where binding `0` is read-only storage and binding `1` is writable storage.
25+
This catches the storage access mismatch class that Naga-only validation can
26+
miss.
27+
- `render-rgba8`: create an RGBA8 render pipeline with explicit vertex and
28+
fragment entry points.
29+
30+
Entry-point flags:
31+
32+
- `--compute-entry NAME`, default `main`
33+
- `--vertex-entry NAME`, default `vs_main`
34+
- `--fragment-entry NAME`, default `fs_main`

tools/wgpu_validation/README.md

Lines changed: 1 addition & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1 @@
1+
README.mbt.md

0 commit comments

Comments
 (0)