facett-core/tests/gpu_taa.rs
//! **GFX_V2 item 7 — TAA really CONVERGES, and stays sharp** (needs `--features wgpu`).
//!
//! The claim is temporal: accumulating N *jittered* frames antialiases an edge that a
//! single frame renders as a staircase. So it is measured, not eyeballed, and the metric
//! is chosen so that the two ways of faking it both fail.
//!
//! **THE METRIC — per-row coverage residual.** A slightly tilted hard edge is rendered
//! and, for each row `y`, the luma is summed across a fixed horizontal band that the
//! edge crosses. That sum is proportional to how much of the band the shape covers, and
//! for a *straight* edge the true coverage is **linear in `y`**. A single aliased frame
//! can only step by whole pixels, so its row sums form a staircase — fit a line and the
//! RMS residual is the aliasing, in luma units. Properly antialiased, the residual
//! collapses.
//!
//! **WHY THAT ALONE IS HOLLOW, AND WHAT CATCHES IT.** A plain blur also flattens the
//! staircase. So the same frame is measured for **transition width** — pixels per row
//! strictly between background and foreground. Real antialiasing of a hard edge gives
//! about one; a blur gives three or more; a dead pass gives zero. The pair
//! (residual collapsed AND width ≈ 1) is satisfied only by genuine sub-pixel
//! convergence.
//!
//! **THE NON-VACUITY THE WHOLE FILE RESTS ON.** The single-frame residual is asserted
//! *large* first. Every "TAA reduced it" assertion is a ratio against a control measured
//! in the same run on the same geometry — never against a constant. And the mutation is
//! **zeroing the jitter**: N identical frames average to the same aliased frame, so the
//! residual does not move. That mutation is what proves the metric measures the jitter
//! rather than the mere presence of a blend.
//!
//! The frames come from the OIT lane rendering an opaque polygon — reusing a pass that
//! already exists rather than adding a test-only rasteriser (LAW #5). Alpha is 1, so the
//! OIT resolve emits a hard-edged shape, which is exactly the aliased input TAA is for.
//!
//! Run: `cargo test -p facett-core --features wgpu --test gpu_taa -- --nocapture`
#![cfg(feature = "wgpu")]
use facett_core::render::gpu::oit::{OitBatch, OitPass};
use facett_core::render::gpu::taa::{jitter_px, TaaPass, TAA_PHASES};
use facett_core::render::gpu::{preferred_backends, read_texture_region, request_best_adapter};
use egui::pos2;
/// The component every row and every SKIP in this file is filed under.
const GATE_COMPONENT: &str = "facett-core";
/// What a skipped arm of this file would have proven, so a SKIPPED matrix row is
/// readable without opening the test.
const GATE_DETAIL: &str = "TAA jitter accumulation on a real device — convergence, sharpness and the no-jitter control";
/// `Rgba8Unorm` throughout, so the readback is the same bytes the resolve wrote and the
/// measurement is not mediated by a float conversion.
const FMT: wgpu::TextureFormat = wgpu::TextureFormat::Rgba8Unorm;
const W: u32 = 128;
const H: u32 = 128;
/// The edge sweeps `EDGE_RUN` px horizontally over the full height — a slope of one
/// pixel every 16 rows, so an aliased render has 16-row-long flat treads and a 255-luma
/// riser between them. Shallow on purpose: a steep edge aliases less and would make the
/// control weak.
const EDGE_X: f32 = 52.0;
const EDGE_RUN: f32 = 8.0;
/// The band the row sums are taken over. Wide enough to contain the edge at every row
/// with slack either side.
const BAND_X0: u32 = 44;
const BAND_X1: u32 = 68;
fn device() -> Option<(wgpu::Adapter, wgpu::Device, wgpu::Queue)> {
let instance = wgpu::Instance::new(wgpu::InstanceDescriptor {
backends: preferred_backends(),
flags: wgpu::InstanceFlags::from_build_config().with_env(),
backend_options: wgpu::BackendOptions::from_env_or_default(),
memory_budget_thresholds: wgpu::MemoryBudgetThresholds::default(),
display: None,
});
let adapter = request_best_adapter(&instance, preferred_backends())?;
let info = adapter.get_info();
// ATTRIBUTION, not decoration: a graceful skip is how a GPU test file goes green
// having run no shader.
eprintln!(
"[gpu_taa] device: {} ({:?}, {:?}) fragment_writable_storage={}",
info.name,
info.backend,
info.device_type,
OitPass::supported(&adapter)
);
let (device, queue) = pollster::block_on(adapter.request_device(&wgpu::DeviceDescriptor {
label: Some("gpu_taa_test"),
required_features: wgpu::Features::empty(),
required_limits: wgpu::Limits::downlevel_defaults(),
..Default::default()
}))
.ok()?;
Some((adapter, device, queue))
}
fn color_target(device: &wgpu::Device) -> (wgpu::Texture, wgpu::TextureView) {
let tex = device.create_texture(&wgpu::TextureDescriptor {
label: Some("gpu_taa_scene"),
size: wgpu::Extent3d { width: W, height: H, depth_or_array_layers: 1 },
mip_level_count: 1,
sample_count: facett_core::render::gpu::NO_MSAA_SAMPLES,
dimension: wgpu::TextureDimension::D2,
format: FMT,
usage: wgpu::TextureUsages::RENDER_ATTACHMENT
| wgpu::TextureUsages::TEXTURE_BINDING
| wgpu::TextureUsages::COPY_SRC,
view_formats: &[],
});
let view = tex.create_view(&Default::default());
(tex, view)
}
/// A white opaque wedge whose right edge is tilted, offset by `(jx, jy)` sub-pixels.
fn wedge(jx: f32, jy: f32) -> OitBatch {
let mut b = OitBatch::new();
let h = H as f32;
let p = |x: f32, y: f32| pos2(x + jx, y + jy);
let white = [1.0, 1.0, 1.0, 1.0];
// (0,0) → (EDGE_X,0) → (EDGE_X+EDGE_RUN,H) → (0,H)
b.push_tri(p(-4.0, -4.0), p(EDGE_X, -4.0), p(EDGE_X + EDGE_RUN, h + 4.0), white, 1.0);
b.push_tri(p(-4.0, -4.0), p(EDGE_X + EDGE_RUN, h + 4.0), p(-4.0, h + 4.0), white, 1.0);
b
}
fn luma(p: &[u8]) -> f64 {
0.2126 * f64::from(p[0]) + 0.7152 * f64::from(p[1]) + 0.0722 * f64::from(p[2])
}
/// Sum of luma across the band at each row — proportional to the covered fraction.
fn row_sums(bytes: &[u8]) -> Vec<f64> {
(0..H)
.map(|y| {
(BAND_X0..BAND_X1)
.map(|x| {
let o = ((y * W + x) * 4) as usize;
luma(&bytes[o..o + 4])
})
.sum()
})
.collect()
}
/// **The aliasing, in luma units.** RMS residual of the row sums about their
/// least-squares line. A straight edge's true coverage is linear in `y`; only
/// quantisation makes it deviate.
fn edge_residual_rms(bytes: &[u8]) -> f64 {
let s = row_sums(bytes);
let n = s.len() as f64;
let mean_x = (s.len() as f64 - 1.0) / 2.0;
let mean_y = s.iter().sum::<f64>() / n;
let (mut sxy, mut sxx) = (0.0, 0.0);
for (i, v) in s.iter().enumerate() {
let dx = i as f64 - mean_x;
sxy += dx * (v - mean_y);
sxx += dx * dx;
}
let slope = if sxx == 0.0 { 0.0 } else { sxy / sxx };
let intercept = mean_y - slope * mean_x;
let sq: f64 = s
.iter()
.enumerate()
.map(|(i, v)| {
let r = v - (slope * i as f64 + intercept);
r * r
})
.sum();
(sq / n).sqrt()
}
/// Mean pixels per row strictly between background and foreground — the edge's softness.
/// 0 = a hard staircase, ~1 = antialiased, >2 = smeared.
fn transition_width(bytes: &[u8]) -> f64 {
let mut total = 0usize;
for y in 0..H {
for x in BAND_X0..BAND_X1 {
let o = ((y * W + x) * 4) as usize;
let l = luma(&bytes[o..o + 4]);
if l > 16.0 && l < 239.0 {
total += 1;
}
}
}
total as f64 / f64::from(H)
}
/// Render one jittered frame of the wedge through the OIT lane into `scene`.
fn render_frame(
device: &wgpu::Device,
queue: &wgpu::Queue,
oit: &mut OitPass,
scene: &wgpu::TextureView,
jx: f32,
jy: f32,
) {
let batch = wedge(jx, jy);
oit.set_frame(queue, [0.0, 0.0, 0.0, 1.0]);
oit.upload(device, queue, &batch);
let mut enc = device.create_command_encoder(&wgpu::CommandEncoderDescriptor {
label: Some("gpu_taa_scene_enc"),
});
oit.record(&mut enc, scene);
queue.submit(Some(enc.finish()));
}
/// IEEE-754 binary16 bits of a normal `f32` — enough to fill an `Rg16Float` velocity
/// texture with a couple of exact constants.
fn f16_bits(v: f32) -> u16 {
if v == 0.0 {
return if v.is_sign_negative() { 0x8000 } else { 0 };
}
let b = v.to_bits();
let sign = ((b >> 31) & 1) as u16;
let exp = ((b >> 23) & 0xFF) as i32 - 127 + 15;
assert!((1..=30).contains(&exp), "f16_bits only handles normals ({v})");
let mant = ((b & 0x007F_FFFF) >> 13) as u16;
(sign << 15) | ((exp as u16) << 10) | (mant & 0x03FF)
}
/// A full-screen constant motion-vector texture, in **UV per frame**.
fn velocity_texture(
device: &wgpu::Device,
queue: &wgpu::Queue,
vx: f32,
vy: f32,
) -> (wgpu::Texture, wgpu::TextureView) {
let tex = device.create_texture(&wgpu::TextureDescriptor {
label: Some("gpu_taa_velocity"),
size: wgpu::Extent3d { width: W, height: H, depth_or_array_layers: 1 },
mip_level_count: 1,
sample_count: facett_core::render::gpu::NO_MSAA_SAMPLES,
dimension: wgpu::TextureDimension::D2,
format: wgpu::TextureFormat::Rg16Float,
usage: wgpu::TextureUsages::TEXTURE_BINDING | wgpu::TextureUsages::COPY_DST,
view_formats: &[],
});
let (bx, by) = (f16_bits(vx), f16_bits(vy));
let mut data = Vec::with_capacity((W * H * 4) as usize);
for _ in 0..(W * H) {
data.extend_from_slice(&bx.to_le_bytes());
data.extend_from_slice(&by.to_le_bytes());
}
queue.write_texture(
wgpu::TexelCopyTextureInfo {
texture: &tex,
mip_level: 0,
origin: wgpu::Origin3d::ZERO,
aspect: wgpu::TextureAspect::All,
},
&data,
wgpu::TexelCopyBufferLayout {
offset: 0,
bytes_per_row: Some(W * 4),
rows_per_image: Some(H),
},
wgpu::Extent3d { width: W, height: H, depth_or_array_layers: 1 },
);
let view = tex.create_view(&Default::default());
(tex, view)
}
/// Accumulate `frames` frames through TAA and read back the resolved image.
///
/// `jitter` selects the sample pattern: `true` = the real Halton phases, `false` = the
/// MUTATION-shaped control that renders every frame at the same position.
fn accumulate(
device: &wgpu::Device,
queue: &wgpu::Queue,
frames: u64,
jitter: bool,
) -> Vec<u8> {
accumulate_with_velocity(device, queue, frames, jitter, None)
}
fn accumulate_with_velocity(
device: &wgpu::Device,
queue: &wgpu::Queue,
frames: u64,
jitter: bool,
velocity: Option<&wgpu::TextureView>,
) -> Vec<u8> {
let (_scene_tex, scene) = color_target(device);
let mut oit = OitPass::new(device, FMT);
oit.ensure(device, W, H, 4);
let mut taa = TaaPass::new(device, FMT);
taa.ensure(device, W, H);
for _ in 0..frames {
let [jx, jy] = if jitter { taa.jitter_px() } else { [0.0, 0.0] };
render_frame(device, queue, &mut oit, &scene, jx, jy);
let mut enc = device.create_command_encoder(&wgpu::CommandEncoderDescriptor {
label: Some("gpu_taa_resolve_enc"),
});
let ok = taa.record(device, queue, &mut enc, &scene, velocity);
assert!(ok, "the TAA pass must actually resolve, not silently decline");
queue.submit(Some(enc.finish()));
}
assert_eq!(taa.frame(), frames, "the pass counted every frame it resolved");
let tex = taa.resolved().expect("a resolved target after accumulating").clone();
read_texture_region(device, queue, &tex, 4, 0, 0, W, H)
}
/// A single unjittered frame straight out of the rasteriser — the aliased control.
fn single_frame(device: &wgpu::Device, queue: &wgpu::Queue) -> Vec<u8> {
let (tex, view) = color_target(device);
let mut oit = OitPass::new(device, FMT);
oit.ensure(device, W, H, 4);
render_frame(device, queue, &mut oit, &view, 0.0, 0.0);
read_texture_region(device, queue, &tex, 4, 0, 0, W, H)
}
const FRAMES: u64 = 64;
#[test]
fn jittered_accumulation_converges_the_edge_and_keeps_it_sharp() {
let Some((adapter, device, queue)) = device() else {
facett_core::testmatrix::gpu_skip(GATE_COMPONENT, "jittered_accumulation_converges_the_edge_and_keeps_it_sharp", "no WebGPU-class adapter", GATE_DETAIL);
return;
};
if !OitPass::supported(&adapter) {
facett_core::testmatrix::gpu_skip(GATE_COMPONENT, "jittered_accumulation_converges_the_edge_and_keeps_it_sharp", "adapter lacks FRAGMENT_WRITABLE_STORAGE", GATE_DETAIL);
return;
}
assert_eq!(FRAMES % u64::from(TAA_PHASES), 0, "accumulate whole jitter cycles");
let one = single_frame(&device, &queue);
let taa = accumulate(&device, &queue, FRAMES, true);
let (r1, rt) = (edge_residual_rms(&one), edge_residual_rms(&taa));
let (w1, wt) = (transition_width(&one), transition_width(&taa));
eprintln!(
"[gpu_taa] single frame: residual={r1:.2} width={w1:.2}px | {FRAMES} jittered frames: \
residual={rt:.2} width={wt:.2}px | reduction={:.2}x",
r1 / rt.max(1e-9)
);
// ── NON-VACUITY: the control is genuinely aliased. ──
// 16-row treads with a 255-luma riser give a staircase residual near 255/sqrt(12).
assert!(
r1 > 40.0,
"the single-frame control must actually be aliased (residual {r1:.2}); if it is not, \
every 'TAA reduced it' assertion below is measuring nothing"
);
assert!(w1 < 0.2, "and hard-edged — a rasterised opaque edge has no partial pixels ({w1:.2}px)");
// The frame is a real picture, not blank: the band must be part covered, part not.
let sums = row_sums(&taa);
let (lo, hi) = sums
.iter()
.fold((f64::MAX, f64::MIN), |(l, h), &v| (l.min(v), h.max(v)));
assert!(
hi - lo > 255.0 * 4.0,
"the wedge must sweep across the band ({lo:.0}..{hi:.0}) — a uniform frame proves nothing"
);
// ── THE HEADLINE: jitter across N frames reduces the edge residual. ──
assert!(
rt < r1 * 0.25,
"{FRAMES} jittered frames must cut the edge residual to under a quarter of one frame's \
({rt:.2} vs {r1:.2})"
);
// ── THE SHARPNESS CONTROL: a blur would also flatten the staircase. ──
assert!(
wt > 0.5,
"the converged edge must actually be soft ({wt:.2}px) — zero partial pixels means the \
resolve produced no sub-pixel detail at all"
);
assert!(
wt < 2.0,
"…but no wider than ~1 px ({wt:.2}px). A box blur reduces the residual too; this is the \
assertion it fails"
);
}
#[test]
fn without_jitter_the_same_accumulation_does_not_converge() {
let Some((adapter, device, queue)) = device() else {
facett_core::testmatrix::gpu_skip(GATE_COMPONENT, "without_jitter_the_same_accumulation_does_not_converge", "no WebGPU-class adapter", GATE_DETAIL);
return;
};
if !OitPass::supported(&adapter) {
facett_core::testmatrix::gpu_skip(GATE_COMPONENT, "without_jitter_the_same_accumulation_does_not_converge", "adapter lacks FRAGMENT_WRITABLE_STORAGE", GATE_DETAIL);
return;
}
// **The mutation, kept as a permanent test.** Same TAA pass, same frame count, same
// geometry — only the sub-pixel offsets are zeroed. Averaging N identical frames is
// still that frame, so if this ever converged, the test above would be measuring the
// blend rather than the jitter.
let one = single_frame(&device, &queue);
let flat = accumulate(&device, &queue, FRAMES, false);
let jittered = accumulate(&device, &queue, FRAMES, true);
let (r1, rf, rj) = (
edge_residual_rms(&one),
edge_residual_rms(&flat),
edge_residual_rms(&jittered),
);
eprintln!(
"[gpu_taa] residual — single {r1:.2} | {FRAMES} UNJITTERED {rf:.2} | {FRAMES} jittered {rj:.2}"
);
assert!(
rf > r1 * 0.9,
"accumulating {FRAMES} IDENTICAL frames must leave the aliasing alone ({rf:.2} vs {r1:.2}) \
— a temporal blend that antialiases without jitter is blurring, not resolving"
);
assert!(
rj < rf * 0.25,
"and the only difference between these two runs is the jitter, so the jitter is what \
converges the edge ({rj:.2} jittered vs {rf:.2} not)"
);
assert!(
transition_width(&flat) < 0.2,
"the unjittered accumulation stays hard-edged, i.e. it really did not antialias"
);
}
#[test]
fn the_first_frame_is_taken_outright_and_a_reset_restarts_the_sequence() {
let Some((adapter, device, queue)) = device() else {
facett_core::testmatrix::gpu_skip(GATE_COMPONENT, "the_first_frame_is_taken_outright_and_a_reset_restarts_the_sequence", "no WebGPU-class adapter", GATE_DETAIL);
return;
};
if !OitPass::supported(&adapter) {
facett_core::testmatrix::gpu_skip(GATE_COMPONENT, "the_first_frame_is_taken_outright_and_a_reset_restarts_the_sequence", "adapter lacks FRAGMENT_WRITABLE_STORAGE", GATE_DETAIL);
return;
}
// Frame 0 has no history, so the resolve must pass the current frame through
// UNCHANGED. If it blended against an uninitialised history, the first frame after
// every teleport would be a dark or garbage flash.
let (scene_tex, scene) = color_target(&device);
let mut oit = OitPass::new(&device, FMT);
oit.ensure(&device, W, H, 4);
let mut taa = TaaPass::new(&device, FMT);
taa.ensure(&device, W, H);
assert_eq!(taa.weight(), 0.0, "no history ⇒ no history weight");
render_frame(&device, &queue, &mut oit, &scene, 0.0, 0.0);
let raw = read_texture_region(&device, &queue, &scene_tex, 4, 0, 0, W, H);
let mut enc = device.create_command_encoder(&Default::default());
assert!(taa.record(&device, &queue, &mut enc, &scene, None));
queue.submit(Some(enc.finish()));
let out = read_texture_region(
&device,
&queue,
taa.resolved().expect("resolved"),
4,
0,
0,
W,
H,
);
assert_eq!(out, raw, "frame 0 must pass through byte-identically");
// After accumulating, a reset must return the pass to that state — same phase, no
// history — otherwise a preset jump reprojects across a discontinuity.
for _ in 0..5 {
let [jx, jy] = taa.jitter_px();
render_frame(&device, &queue, &mut oit, &scene, jx, jy);
let mut enc = device.create_command_encoder(&Default::default());
taa.record(&device, &queue, &mut enc, &scene, None);
queue.submit(Some(enc.finish()));
}
assert_eq!(taa.frame(), 6);
assert!(taa.weight() > 0.5, "the ramp is well under way");
assert_ne!(taa.jitter_px(), jitter_px(0), "…at a later phase");
taa.reset();
assert_eq!(taa.frame(), 0);
assert_eq!(taa.weight(), 0.0, "a reset drops the history weight");
assert_eq!(taa.jitter_px(), jitter_px(0), "and restarts the jitter sequence");
render_frame(&device, &queue, &mut oit, &scene, 0.0, 0.0);
let mut enc = device.create_command_encoder(&Default::default());
taa.record(&device, &queue, &mut enc, &scene, None);
queue.submit(Some(enc.finish()));
let after = read_texture_region(
&device,
&queue,
taa.resolved().expect("resolved"),
4,
0,
0,
W,
H,
);
assert_eq!(
after, raw,
"and the first frame after a reset is again taken outright — no stale history bleeds in"
);
}
#[test]
fn a_history_whose_reprojection_leaves_the_frame_cannot_converge_the_edge() {
let Some((adapter, device, queue)) = device() else {
facett_core::testmatrix::gpu_skip(GATE_COMPONENT, "a_history_whose_reprojection_leaves_the_frame_cannot_converge_the_edge", "no WebGPU-class adapter", GATE_DETAIL);
return;
};
if !OitPass::supported(&adapter) {
facett_core::testmatrix::gpu_skip(GATE_COMPONENT, "a_history_whose_reprojection_leaves_the_frame_cannot_converge_the_edge", "adapter lacks FRAGMENT_WRITABLE_STORAGE", GATE_DETAIL);
return;
}
// **The velocity path, exercised.** With zero velocity `huv` never leaves `[0,1]` and
// the reprojection is the identity, so the tests above say nothing about motion at
// all. A velocity of 2 UV per frame puts EVERY pixel's reprojection outside the
// frame: no usable history anywhere, so the resolve must fall back to the current
// frame and therefore must NOT converge, however many frames it accumulates.
//
// WHAT THIS DOES AND DOES NOT PROVE, measured. It proves the rejection happens. It
// does NOT isolate the explicit bounds test in the shader: deleting that test moves
// the residual below from 71.70 to 71.28 and nothing here notices, because the 3x3
// box clamp already bounds an off-screen sample into the current frame's own range.
// So this asserts the REJECTION, and the bounds test is documented as belt on braces
// in `taa.wgsl` rather than claimed to be under test.
let (_vt, vel) = velocity_texture(&device, &queue, 2.0, 2.0);
let still = accumulate(&device, &queue, FRAMES, true);
let flung = accumulate_with_velocity(&device, &queue, FRAMES, true, Some(&vel));
let (rs, rf) = (edge_residual_rms(&still), edge_residual_rms(&flung));
let one = edge_residual_rms(&single_frame(&device, &queue));
eprintln!(
"[gpu_taa] residual — single {one:.2} | {FRAMES} frames, no motion {rs:.2} | \
{FRAMES} frames, all reprojections OFF SCREEN {rf:.2}"
);
assert!(
rs < one * 0.25,
"the zero-velocity control must still converge ({rs:.2} vs {one:.2}) — without it, a \
resolve that was simply broken would satisfy the assertion below"
);
assert!(
rf > one * 0.9,
"every reprojection lands off screen, so no history is usable and the aliasing must \
survive ({rf:.2} vs a single frame's {one:.2})"
);
assert!(
transition_width(&flung) < 0.2,
"…and the edge must stay hard: a rejected history contributes nothing, it does not \
contribute a clamped border texel"
);
}