Your browser sucks, JavaScript disabled! Gunnar's console needs JavaScript and WebAssembly. Git over SSH on port 2222 does not.

nordisk/facett

facett-map/tests/gpu_lane_paint_order.rs

//! **THE DECLARED PAINT ORDER, ASSERTED ON THE LANE THAT SHIPS.**
//!
//! # Why this file exists next to `paint_order_pixels.rs`
//!
//! `tests/paint_order_pixels.rs` proves the declared order on **pixels** and was proven
//! red — and every one of its pixel claims runs on the **CPU painter**. Its own doc says
//! so: *"a headless `Harness` installs no `RenderState`, so `renderer_installed()` is
//! false and the frame never places a callback"*. So on 2026-08-22 the order was fixed
//! and guarded for the lane korp does **not** use on a box with a discrete GPU, and the
//! lane it **does** use was covered only by a *bake-stream* assertion
//! (`the_gpu_bake_emits_the_declared_order_whatever_the_input_order`) — which reads the
//! vertex buffer, not the framebuffer, and therefore cannot see the one thing that
//! decides the stack on the device: **the order the render pass submits its draws in.**
//!
//! The owner on 2026-08-23, after that fix shipped: *"i still see the bug with layers
//! overwriting itself"*. This file is the reproduction.
//!
//! # The defect, in one line
//!
//! `MapPaintCallback::paint` (facett-map/src/gpu/mod.rs:1902 tile lane, :1955 legacy
//! lane) submits **THREE HARD BUCKETS — all fills, then all fill outlines, then all
//! lines — with no depth buffer and no `PaintKey` between them.** `PaintKey` orders the
//! stream *inside* a bucket and cannot reach across one. So on the GPU lane:
//!
//! * every LINE paints over every FILL, whatever `layer`/`bridge`/`tunnel` say;
//! * a `tunnel=yes` road under a lake paints **on top of the lake**;
//! * and `paint_order_pixels::a_bridge_paints_over_the_water_and_a_tunnel_under_it`
//!   stays green while it happens, because it never reaches the device lane.
//!
//! # The two claims
//!
//! 1. **`a_bridge_over_the_lake_shows_the_ROAD_on_the_gpu_lane`** — the control. It
//!    passes before and after any fix, and it is what makes claim 2 non-vacuous: it
//!    proves the road, the lake and the crossing pixel are all really there and that the
//!    GPU lane really drew them.
//! 2. **`a_tunnel_under_the_lake_shows_the_WATER_on_the_gpu_lane`** — the defect. RED as
//!    of 2026-08-23 on oden / RTX 4090: the crossing pixel is the road, not the water.
//!
//! Both are asserted on the framebuffer, and both refuse to pass unless
//! `lane::lane_use().frames()` grew — a frame the CPU painter drew is not evidence about
//! the GPU lane, and that substitution is exactly how this hole stayed open.
//!
//! Run (needs the real device, opens NO window):
//! ```text
//! env -u WAYLAND_DISPLAY -u DISPLAY FACETT_TEST_GPU=discrete \
//!   cargo test -p facett-map --features wgpu --test gpu_lane_paint_order -- --nocapture
//! ```
#![cfg(feature = "wgpu")]

use facett_core::render::lane;
use facett_map::layer::{LatLon, MapFacet, MapTheme, MapView, WaysLayer};
use facett_map::{Way, WayKind};

const CANVAS: egui::Vec2 = egui::vec2(520.0, 400.0);

/// **The shared one-device-at-a-time lock** — `common/one_gpu.rs`, the SAME static every
/// other rendering file in this directory takes. This file used to declare its own
/// `static … Mutex<()>`; five files declared a copy of the same two lines, which is the
/// twin LAW 5 forbids — and a twinned LOCK is worse than a twinned helper, because two
/// mutexes over one device serialise nothing the moment a file reaches the device
/// through the other one's name.
#[path = "common/one_gpu.rs"]
mod one_gpu;

/// **The device opt-in, shared** — `common/gpu_gate.rs`, lane GG #5. See that module for
/// why the vocabulary grew a word for an integrated GPU.
#[path = "common/gpu_gate.rs"]
mod gpu_gate;

/// A closed axis-aligned ring, built through the SAME constructor korp's served and
/// tiled paths use, so the tags here are parsed by `paint_order::parse_osm_z`.
fn ring(id: i64, lon: f64, lat: f64, w: f64, tags: &[(&str, &str)]) -> Way {
    let t: Vec<(String, String)> = tags.iter().map(|(k, v)| ((*k).to_string(), (*v).to_string())).collect();
    let r = vec![
        (lon - w, lat - w),
        (lon + w, lat - w),
        (lon + w, lat + w),
        (lon - w, lat + w),
        (lon - w, lat - w),
    ];
    Way::from_lonlat_with(id, &r, &t, &facett_map::MapPalette::cartographic(false))
        .expect("a 5-vertex closed ring is a way")
}

/// A straight two-point line at latitude `lat`.
fn line(id: i64, lon0: f64, lon1: f64, lat: f64, tags: &[(&str, &str)]) -> Way {
    let t: Vec<(String, String)> = tags.iter().map(|(k, v)| ((*k).to_string(), (*v).to_string())).collect();
    Way::from_lonlat_with(id, &[(lon0, lat), (lon1, lat)], &t, &facett_map::MapPalette::cartographic(false))
        .expect("a 2-vertex line is a way")
}

/// A fresh real `RenderState` with `GpuMapRenderer` installed — exactly what korp's
/// eframe creation closure does, and deliberately NOT
/// `egui_kittest::wgpu::default_wgpu_setup()`, whose selector prefers software
/// rasterisers (see `tests/map_facet_gpu_lane.rs` for the full argument).
fn fresh_render_state() -> egui_wgpu::RenderState {
    let setup = facett_core::render::gpu::adapter_wgpu::facett_wgpu_options().wgpu_setup;
    let rs = egui_kittest::wgpu::create_render_state(setup, Default::default());
    facett_map::gpu::install_renderer(&rs);
    rs
}

/// Drive one facet through the real device and hand back
/// `(lane frames encoded, framebuffer)`.
///
/// Every step is rendered, not just the last: the upload-once latch puts the geometry
/// payload on the FIRST frame's callback, so a `run_steps(n); render()` shape submits
/// only the last frame — whose callback carries `geometry: None` — and the lane sits
/// legitimately at 0. Same reasoning, same shape, as `map_facet_gpu_lane::drive`.
fn drive(mut f: MapFacet, steps: u32) -> (u64, image::RgbaImage) {
    use egui_kittest::wgpu::WgpuTestRenderer;
    let _turn = one_gpu::hold();
    let before = lane::lane_use().frames();
    let mut h = egui_kittest::Harness::builder()
        .with_size(CANVAS)
        .renderer(WgpuTestRenderer::from_render_state(fresh_render_state()))
        .build_ui(move |ui| f.ui(ui));
    let mut img = None;
    for _ in 0..steps {
        img = Some(h.render().expect("the wgpu backend produced a frame"));
        h.step();
    }
    (lane::lane_use().frames() - before, img.expect("at least one frame"))
}

/// The rendered colour at a fraction of the pane.
fn px_at(img: &image::RgbaImage, fx: f32, fy: f32) -> [u8; 4] {
    let x = ((fx * img.width() as f32) as u32).min(img.width() - 1);
    let y = ((fy * img.height() as f32) as u32).min(img.height() - 1);
    img.get_pixel(x, y).0
}

fn near(a: [u8; 4], b: [u8; 4], tol: i32) -> bool {
    (0..3).all(|i| (i32::from(a[i]) - i32::from(b[i])).abs() <= tol)
}

/// **Is `px` a mix of `a` and `b`** — within `tol` per channel of some point on the RGB
/// segment `a → b`, ends included? The dash lane (HO-T3) paints a rail's centreline in
/// two inks, core and casing, and feathers the dash END with the same ramp as the side, so
/// a pixel straddling a dash edge is a convex blend of the two. Such a pixel is the rail's
/// as much as a pure one is. The mix `t` is read off the channel with the widest span, so
/// two near-equal channels cannot make the division meaningless.
fn on_blend(px: [u8; 4], a: [u8; 4], b: [u8; 4], tol: i32) -> bool {
    let span = |i: usize| f32::from(b[i]) - f32::from(a[i]);
    let widest = (0..3).max_by(|&i, &j| span(i).abs().total_cmp(&span(j).abs())).expect("three channels");
    if span(widest).abs() < 1.0 {
        return near(px, a, tol);
    }
    let t = ((f32::from(px[widest]) - f32::from(a[widest])) / span(widest)).clamp(0.0, 1.0);
    (0..3).all(|i| {
        let at_t = f32::from(a[i]) + t * span(i);
        (f32::from(px[i]) - at_t).abs() <= tol as f32
    })
}

/// The fixture: a wide lake with one motorway straight across its middle, `tags` on the
/// road. The lake is fed LAST so "whoever was read last wins" would put water on the
/// crossing every time — the road has to win on its declared rank, not on its position.
fn lake_and_road(tags: &[(&str, &str)]) -> MapFacet {
    let lake = ring(1, 18.060, 59.330, 0.020, &[("natural", "water")]);
    let road = line(2, 18.040, 18.080, 59.330, tags);
    assert_eq!(lake.style.kind, WayKind::Water, "fixture: the ring really classifies as water");
    assert_eq!(road.style.kind, WayKind::Road, "fixture: the line really classifies as a road");
    assert!(lake.fill.is_some(), "fixture: the lake must be a FILL, or there is no bucket to cross");
    assert!(road.fill.is_none(), "fixture: the road must be a LINE, or there is no bucket to cross");
    let view = MapView::new(LatLon::new(59.33, 18.06), 13.0, (CANVAS.x, CANVAS.y));
    MapFacet::new("gpu-paint-order", view, MapTheme::default())
        .with_layer(Box::new(WaysLayer::new("osm-ways", vec![road, lake])))
}

/// `true` when this host is allowed to answer, and the skip is filed as a NON-passing
/// matrix row rather than as a silent `ok`.
///
/// The body moved to `common/gpu_gate.rs` in lane GG #5: three files carried a copy and
/// two of them recognised only `discrete|software`, which left the owner's **integrated**
/// Radeon 890M — the box that shows every defect this file is about — with no truthful
/// value to set.
fn gated(name: &str) -> bool {
    gpu_gate::gated("facett-map", name)
}

/// The fixture for the **TILE lane**: one building and one lake that overlap, each
/// stamped with a `tile_key` so `add_gpu_tiled_callback` groups them into two separate
/// tiles. `swap` puts the lake in the low-numbered tile instead of the high one.
///
/// The declared order is unambiguous and has nothing to do with tiles: water (`rank`
/// `AREA_BAND + 100`) paints UNDER a building (`AREA_BAND + 300`), so the building must
/// be visible in both arrangements.
fn two_tiles(swap: bool) -> MapFacet {
    // A big lake and a small building inside it, overlapping by construction.
    let mut lake = ring(1, 18.060, 59.330, 0.020, &[("natural", "water")]);
    let mut building = ring(2, 18.060, 59.330, 0.006, &[("building", "yes")]);
    assert_eq!(lake.style.kind, WayKind::Water, "fixture: the big ring is water");
    assert_eq!(building.style.kind, WayKind::Building, "fixture: the small ring is a building");
    assert!(
        facett_map::way_rank(&lake) < facett_map::way_rank(&building),
        "fixture: the DECLARED order already says water ({}) under buildings ({}) — this test \
         is about whether the device honours it, not about what it should be",
        facett_map::way_rank(&lake),
        facett_map::way_rank(&building)
    );
    // Two different tiles. `add_gpu_tiled_callback` sorts the groups by `tile_key`, so
    // tile 1 is submitted before tile 2 whichever way round the content is.
    lake.tile_key = if swap { 1 } else { 2 };
    building.tile_key = if swap { 2 } else { 1 };
    let view = MapView::new(LatLon::new(59.33, 18.06), 13.0, (CANVAS.x, CANVAS.y));
    MapFacet::new("gpu-tile-order", view, MapTheme::default())
        .with_layer(Box::new(WaysLayer::new("osm-ways", vec![lake, building])))
}

/// The three reference colours of one render, plus the frame count that proves the GPU
/// lane — not the CPU painter — produced them.
struct Look {
    frames: u64,
    /// Pane centre: the middle of the lake, on the road's line.
    crossing: [u8; 4],
    /// Inside the lake, well off the road.
    open_water: [u8; 4],
    /// Outside the lake entirely.
    /// Outside the lake AND clear of every pane overlay — see [`look`].
    background: [u8; 4],
    /// A SECOND background sample, far from the first. If an overlay ever creeps over
    /// either of them these two disagree and the fixture says so, instead of quietly
    /// comparing the map against a chip.
    background_alt: [u8; 4],
    distinct: usize,
}

fn look(tags: &[(&str, &str)], shot: &str) -> Look {
    let (frames, img) = drive(lake_and_road(tags), 3);
    let path = format!("{}/scratch/cargo/gpu_lane_paint_order_{shot}.png", std::env::var("HOME").unwrap_or_default());
    if image::DynamicImage::ImageRgba8(img.clone()).save(&path).is_ok() {
        eprintln!("LOOK: saved {path} ({}×{})", img.width(), img.height());
    }
    let distinct: std::collections::HashSet<[u8; 4]> = img.pixels().map(|p| p.0).collect();
    Look {
        frames,
        crossing: px_at(&img, 0.5, 0.5),
        open_water: px_at(&img, 0.5, 0.30),
        // ── THE BACKGROUND PROBE MOVED, and the honest reason ──────────────────────
        // It was (0.02, 0.02) — pixel (10, 8) on this 520x400 pane. On 2026-08-23 the
        // lane-flag chip landed in that corner
        // (`facett_core::render::lane::paint_lane_flag`, an OPAQUE plate at
        // `rect.left() + 4, rect.top() + 4`).
        //
        // MEASURED off the saved framebuffer rather than reasoned about, because the first
        // reasoning was WRONG: the chip occupies y = 10..25 and the probe sat at y = 8, so
        // it was still reading real paper (31, 31, 38) and the guard was never actually
        // compromised. It survived by TWO PIXELS.
        //
        // Which is the point of moving it anyway. A probe two pixels from an opaque overlay
        // is not measuring the background on purpose, it is doing so by luck, and the next
        // change to this `Ui`'s padding collects the debt silently — the chip's plate
        // (10, 22, 14) is just as far from the water (18, 58, 90) as the paper is, so every
        // assertion below would have stayed GREEN while measuring the wrong pixels.
        //
        // Bottom-right and bottom-left: outside the lake (roughly x 144..376, y 84..316 at
        // this camera) and clear of the three corners overlays claim — the lane flag takes
        // top-left, `devid::badge` and the znippy logo take top-right.
        background: px_at(&img, 0.95, 0.95),
        background_alt: px_at(&img, 0.05, 0.95),
        distinct: distinct.len(),
    }
}

/// Everything that has to be true before either claim means anything: the GPU lane
/// really encoded frames, and the lake really painted water where no road is.
fn assert_the_fixture_is_alive(l: &Look, what: &str) {
    assert!(
        l.frames > 0,
        "{what}: the GPU lane encoded ZERO frames, so this render is the CPU painter and \
         says nothing about the device lane — which is exactly the substitution that let \
         the three-bucket collapse hide behind a green `paint_order_pixels`"
    );
    assert!(
        l.distinct >= 3,
        "{what}: the render has only {} distinct colours — background + water + road is \
         the minimum, or an 'identical' pixel proves nothing (the identity trap)",
        l.distinct
    );
    // The two background probes must AGREE. They are far apart and both meant to be bare
    // paper; a disagreement means an overlay (or the lake) has crept over one of them and
    // the "background" this fixture compares against is no longer the background.
    assert_eq!(
        l.background, l.background_alt,
        "{what}: the two background probes disagree ({:?} at bottom-right vs {:?} at \
         bottom-left) — something is painting over one of them, so every comparison against \
         'the background' below is measuring the wrong pixels",
        l.background, l.background_alt
    );
    assert!(
        !near(l.open_water, l.background, 6),
        "{what}: open water {:?} is indistinguishable from the background {:?} — the lake \
         did not paint, so no ordering claim about it can be tested",
        l.open_water,
        l.background
    );
}

// ─────────────────────────────────────────────────────────────────────────────

/// **THE CONTROL.** A `bridge=yes` motorway over the lake must show the ROAD at the
/// crossing. This is expected GREEN before and after any fix: a line over a fill is what
/// the three-bucket order already does, and it is also what the declared order wants.
///
/// It is here because claim 2 below is an assertion about a *colour that is already on
/// screen somewhere*, and without this control a fix that simply stopped drawing roads
/// would turn claim 2 green.
#[test]
fn a_bridge_over_the_lake_shows_the_road_on_the_gpu_lane() {
    if !gated("a_bridge_over_the_lake_shows_the_road_on_the_gpu_lane") {
        return;
    }
    let l = look(&[("highway", "motorway"), ("bridge", "yes")], "bridge");
    assert_the_fixture_is_alive(&l, "bridge");
    let ok = !near(l.crossing, l.open_water, 6);
    nornir_testmatrix::functional_status(
        "facett-map",
        "gpu_lane_bridge_over_water",
        ok,
        &format!(
            "{} lane frames; crossing {:?}, open water {:?}, background {:?}",
            l.frames, l.crossing, l.open_water, l.background
        ),
    );
    assert!(
        ok,
        "a BRIDGE over the lake must show the road on the crossing pixel: got {:?}, open \
         water is {:?}",
        l.crossing, l.open_water
    );
}

/// **THE DEFECT.** A `tunnel=yes` motorway under the lake must leave the WATER's own
/// colour on the crossing pixel. `parse_osm_z` reads the tag, `way_rank` sinks the way a
/// whole `LAYER_STEP` below the lake, and the bake sorts the stream on it — and none of
/// that can help, because the render pass draws **all fills, then all outlines, then all
/// lines** and the road is a line.
///
/// RED as of 2026-08-23 (oden, RTX 4090): the crossing pixel is the road. The CPU-painter
/// twin of this claim (`paint_order_pixels::a_bridge_paints_over_the_water_and_a_tunnel_under_it`)
/// is GREEN on the same tree — which is the whole reason this file exists.
///
/// The cure this asks for is a submission order keyed on
/// `PaintKey`/`OsmZ::effective_layer`, not three fixed buckets: draw
/// `fills(layer) → outlines(layer) → lines(layer)` per OSM layer ascending (the
/// osm-carto rule), or give the lane a depth attachment and put the rank in Z.
#[test]
fn a_tunnel_under_the_lake_shows_the_water_on_the_gpu_lane() {
    if !gated("a_tunnel_under_the_lake_shows_the_water_on_the_gpu_lane") {
        return;
    }
    let l = look(&[("highway", "motorway"), ("tunnel", "yes")], "tunnel");
    assert_the_fixture_is_alive(&l, "tunnel");
    let ok = near(l.crossing, l.open_water, 6);
    nornir_testmatrix::functional_status(
        "facett-map",
        "gpu_lane_tunnel_under_water",
        ok,
        &format!(
            "{} lane frames; crossing {:?}, open water {:?}, background {:?}",
            l.frames, l.crossing, l.open_water, l.background
        ),
    );
    assert!(
        ok,
        "a TUNNEL under the lake must leave the WATER on the crossing pixel: got {:?}, \
         open water is {:?}. The GPU lane draws all fills, then all outlines, then all \
         lines (gpu/mod.rs:1902 tile, :1955 legacy) — so no `layer`/`tunnel` rank can put \
         a line under a fill, and `PaintKey` never gets a say across the bucket boundary.",
        l.crossing, l.open_water
    );
}

/// **WHICH TILE A FEATURE LANDED IN MUST NOT DECIDE WHAT COVERS WHAT.**
///
/// This is the reported symptom stated as a property, on the axis that carries it. The
/// TILE lane (`MapFacet::add_gpu_tiled_callback`, layer.rs:3247 — the lane korp runs
/// against real served tiles) groups by `Way::tile_key`, and
/// `MapPaintCallback::paint` then walks `renderer.tile_current` **once per bucket**:
/// every tile's fills, then every tile's outlines, then every tile's lines. Each tile's
/// stream is sorted by `PaintKey` *independently*, so across tiles the stack is decided
/// by `tile_key` order and not by class.
///
/// The code says so itself, at layer.rs:3273: *"same-class overlaps across tiles are
/// rare and class order is preserved across ALL tiles by the renderer's three passes"*.
/// Class order is preserved across tiles only for the same class. A **building in tile 1
/// and a lake in tile 2** is a different-class cross-tile overlap, and there the lake —
/// which the declared order puts underneath — is submitted second and covers the
/// building.
///
/// The oracle is the pair, not a single pixel: the same two features in the two possible
/// tile assignments must render identically. A single-render assertion could be argued
/// away as a fixture choice; a pair that disagrees can only be the tile order deciding
/// the paint order. That is "the cache decides the drawing" made reproducible.
///
/// RED as of 2026-08-23 (oden, RTX 4090).
#[test]
fn which_tile_a_feature_landed_in_cannot_decide_what_covers_what() {
    if !gated("which_tile_a_feature_landed_in_cannot_decide_what_covers_what") {
        return;
    }
    let shot = |swap: bool, name: &str| {
        let (frames, img) = drive(two_tiles(swap), 3);
        let path = format!("{}/scratch/cargo/gpu_lane_paint_order_{name}.png", std::env::var("HOME").unwrap_or_default());
        if image::DynamicImage::ImageRgba8(img.clone()).save(&path).is_ok() {
            eprintln!("LOOK: saved {path} ({}×{})", img.width(), img.height());
        }
        (frames, img)
    };
    let (f_a, a) = shot(false, "tile_building_first");
    let (f_b, b) = shot(true, "tile_lake_first");

    // NON-VACUITY, three ways. Without all three an "identical" pair proves nothing.
    assert!(
        f_a > 0 && f_b > 0,
        "both renders must come off the GPU lane (frames {f_a} / {f_b}); a CPU-painter \
         frame says nothing about the device submission order"
    );
    assert_eq!((a.width(), a.height()), (b.width(), b.height()), "same pane both ways");
    let distinct: std::collections::HashSet<[u8; 4]> = a.pixels().map(|p| p.0).collect();
    assert!(
        distinct.len() >= 3,
        "the fixture must render background + lake + building ({} distinct colours) — two \
         blank panes are also identical",
        distinct.len()
    );
    // The building's own ink must be ON SCREEN in at least one of the two arrangements,
    // or the test would pass on a lane that lost the building entirely.
    let centre_a = px_at(&a, 0.5, 0.5);
    let centre_b = px_at(&b, 0.5, 0.5);
    let water_a = px_at(&a, 0.5, 0.30);
    assert!(
        !near(centre_a, water_a, 6) || !near(centre_b, water_a, 6),
        "the building must be visible over the lake in at least one arrangement — centres \
         {centre_a:?} / {centre_b:?} are both the water colour {water_a:?}, so the building \
         is simply missing and no ordering claim is being tested"
    );

    let diff = a.pixels().zip(b.pixels()).filter(|(x, y)| x.0 != y.0).count();
    let total = (a.width() * a.height()) as usize;
    nornir_testmatrix::functional_status(
        "facett-map",
        "gpu_tile_lane_order_is_tile_independent",
        diff == 0,
        &format!(
            "{diff} of {total} pixels differ; centres {centre_a:?} vs {centre_b:?}, water {water_a:?}"
        ),
    );
    assert_eq!(
        diff, 0,
        "moving the SAME two features between two tiles changed {diff} of {total} pixels \
         ({:.2} %). The declared order says water under buildings whatever tile either sat \
         in; the tile lane submits fills tile-by-tile (gpu/mod.rs:1904) so tile 2's lake \
         covers tile 1's building. Centres: building-in-tile-1 {centre_a:?}, \
         lake-in-tile-1 {centre_b:?}, open water {water_a:?}.",
        100.0 * diff as f64 / total as f64
    );
}

/// **THE PRICE OF THE BAND-MAJOR ORDER, MEASURED.**
///
/// The cure splits every tile into one group per draw band, and a group is one VRAM
/// entry, one cull dispatch and up to three `draw_indirect`s. A correct order bought at
/// 4× the draw calls is not obviously a win, so the multiplier is measured here rather
/// than argued about — and bounded, so the two ways it could blow up go red:
///
/// * splitting on the full `PaintKey::rank` instead of the band would add the `Lod` tier
///   to the key and multiply the groups by up to four;
/// * a future class ladder that grows without anyone noticing would show up as a rising
///   groups-per-tile ratio.
///
/// The fixture is the real class mix at a city zoom — water, park and building fills,
/// coastline, path and road lines, one bridge and one tunnel — spread over 8 tiles, i.e.
/// exactly the shape a served z12 viewport hands the lane.
///
/// MEASURED, oden, 2026-08-23: the numbers this prints are the ones quoted in the report.
///
/// CPU-only: it counts what the facet would hand over, so it needs no device and no gate.
#[test]
fn the_band_major_group_count_stays_within_a_bounded_multiple_of_the_tile_count() {
    const TILES: u64 = 8;
    let pal = facett_map::MapPalette::cartographic(false);
    // (tags, closed?) — the ladder `style.rs` actually classifies, both bands.
    let kinds: &[(&[(&str, &str)], bool)] = &[
        (&[("natural", "water")], true),
        (&[("leisure", "park")], true),
        (&[("building", "yes")], true),
        (&[("landuse", "industrial")], true),
        (&[("natural", "coastline")], false),
        (&[("waterway", "river")], false),
        (&[("highway", "footway")], false),
        (&[("highway", "motorway")], false),
        (&[("highway", "residential")], false),
        (&[("highway", "motorway"), ("bridge", "yes")], false),
        (&[("highway", "motorway"), ("tunnel", "yes")], false),
    ];
    let mut ways = Vec::new();
    let mut id = 1i64;
    for tile in 1..=TILES {
        for (k, closed) in kinds {
            let lon = 18.0 + (tile as f64) * 0.01;
            let mut w = if *closed {
                ring(id, lon, 59.33, 0.002, k)
            } else {
                line(id, lon, lon + 0.002, 59.33, k)
            };
            w.tile_key = tile;
            ways.push(w);
            id += 1;
            let _ = &pal;
        }
    }
    let view = MapView::new(LatLon::new(59.33, 18.06), 12.0, (CANVAS.x, CANVAS.y));
    let f = MapFacet::new("census", view, MapTheme::default())
        .with_layer(Box::new(WaysLayer::new("osm-ways", ways)));
    let (groups, tiles) = f.gpu_group_census_for_test(&[0]);
    let per_tile = groups as f64 / tiles as f64;
    eprintln!(
        "LOOK: band-major census — {groups} groups over {tiles} tiles = {per_tile:.2} groups/tile \
         ({} feature classes in the fixture)",
        kinds.len()
    );
    assert_eq!(tiles as u64, TILES, "the fixture must really span {TILES} tiles");
    // NON-VACUOUS: more than one group per tile, or the band never entered the key and
    // this test is measuring the bug it was written to bound.
    assert!(
        per_tile > 1.0,
        "the census found {per_tile:.2} groups per tile — the band is not in the group key \
         and the order this file proves is not actually being produced"
    );
    // The CEILING. A band is (osm layer, class); the fixture has 11 classes across two
    // layers plus a bridge and a tunnel, so 11 is the honest maximum and anything past
    // it means the tier (or something finer) leaked into the key.
    assert!(
        per_tile <= kinds.len() as f64,
        "{per_tile:.2} groups per tile is more than the {} distinct feature classes in the \
         fixture — something finer than the draw band is splitting the groups, and every \
         extra group is a VRAM entry, a cull dispatch and up to three draws per frame",
        kinds.len()
    );
    nornir_testmatrix::functional_status(
        "facett-map",
        "band_major_group_census",
        per_tile > 1.0 && per_tile <= kinds.len() as f64,
        &format!("{groups} groups / {tiles} tiles = {per_tile:.2} per tile, {} classes", kinds.len()),
    );
}

// ─────────────────────────────────────────────────────────────────────────────
// THE LANDMASS FLOOR — "vita marken tar över allt" (owner, 2026-08-30)
// ─────────────────────────────────────────────────────────────────────────────

/// **THE WHITE LAND.** The day cartographic paper `#F2EFE9` — the colour korp hands
/// `LandmassLayer::from_rings` (`korp/src/map_views.rs:234`, `theme(..).palette.land`)
/// and the one the owner means by *"vita marken"*. Named here so the fixture's white is
/// the shipped white and not a test-only colour, and so it stays legible against this
/// file's night-palette ways.
const LAND_WHITE: egui::Color32 = egui::Color32::from_rgb(0xF2, 0xEF, 0xE9);

/// The fixture for the **landmass floor**: a white land silhouette covering the whole
/// pane, a lake in the middle of it, and one motorway that is either an ordinary
/// surface road or a `tunnel=yes` — **the only difference between the two arms.**
///
/// Same geometry, same colours, same layer order, same camera. If the two renders
/// disagree, the disagreement can only come from the `tunnel` tag, i.e. from *what has
/// been loaded* — which is the reported symptom stated as a property.
///
/// MEASURED, why a tunnel is a realistic thing to have in the loaded set: on a real
/// Sweden pyramid tile (`z9/288/125.parquet`, 2026-08-31) **11 of 1 493 ways (0.74 %)**
/// carry `tunnel=*` and 11 carry `layer < 0`. The lane's rank floor is scanned over
/// EVERY way of every GPU layer with no viewport cull, so one such way anywhere in the
/// served set is enough.
fn land_lake_and_road(tunnel: bool) -> MapFacet {
    use facett_map::layer::LandmassLayer;
    let lake = ring(1, 18.060, 59.330, 0.020, &[("natural", "water")]);
    assert!(lake.fill.is_some(), "fixture: the lake must be a FILL");
    let road_tags: &[(&str, &str)] = if tunnel {
        &[("highway", "motorway"), ("tunnel", "yes")]
    } else {
        &[("highway", "motorway")]
    };
    // Well below the lake, so it never touches the probe pixels: this way is here to be
    // LOADED, not to be looked at.
    let road = line(2, 18.030, 18.090, 59.3150, road_tags);
    assert!(road.fill.is_none(), "fixture: the road must be a LINE");
    // A land silhouette far larger than the pane — the country polygon korp pushes
    // through `LandmassLayer::from_rings(.., palette.land)` (korp/src/map_views.rs:234).
    let land = vec![vec![
        (17.50, 59.10),
        (18.60, 59.10),
        (18.60, 59.60),
        (17.50, 59.60),
        (17.50, 59.10),
    ]];
    let view = MapView::new(LatLon::new(59.33, 18.06), 13.0, (CANVAS.x, CANVAS.y));
    MapFacet::new("landmass-floor", view, MapTheme::default())
        .with_layer(Box::new(LandmassLayer::from_rings("landmass", land, LAND_WHITE)))
        .with_layer(Box::new(WaysLayer::new("osm-ways", vec![road, lake])))
}

/// **THE WHITE LAND MUST STAY UNDER THE MAP, WHATEVER IS LOADED.**
///
/// # What the owner reported (2026-08-30)
///
/// *"vita marken tar över allt"* — the white land paints over everything, and it VARIES
/// with what has been loaded.
///
/// # The mechanism, and why it is arithmetic and not a race
///
/// `MapFacet::ui` inserts the GPU callback as ONE opaque slot at the lane's own rank
/// FLOOR (`gpu_lane_rank_floor`, the minimum `paint_order::way_rank` over every way the
/// lane draws). Every CPU feature whose rank is **below** that floor paints before the
/// slot; everything else paints after it.
///
/// `way_rank` is `effective_layer * LAYER_STEP + band + class + tier`, and `LAYER_STEP`
/// is 100 000. A single `tunnel=yes` / `layer=-1` way therefore has a rank of about
/// **−98 800** — while [`facett_map::paint_order::LANDMASS_BAND`] was **−1 000**. So the
/// landmass, which the constant's own doc calls *"the floor, under every other band"*,
/// was NOT under the floor at all: one tunnel anywhere in the loaded set dragged the
/// lane's slot below it, the landmass fell into the CPU **tail**, and the white
/// silhouette painted over the entire GPU-rendered map.
///
/// That is every part of the report: it covers everything (the silhouette is the whole
/// country), and it varies with what has been loaded (whether a negative-layer way is in
/// the set).
///
/// # The oracle
///
/// **Framebuffer pixels off the real device lane**, in a PAIR. The two arms differ by
/// exactly one OSM tag; the control (`tunnel = false`) must show water at the pane
/// centre, and so must the tunnel arm. A single render could be argued away as a
/// fixture choice — a pair that disagrees can only be the loaded set deciding the stack.
///
/// # Proven RED
///
/// With `LANDMASS_BAND = -1_000` the tunnel arm's centre pixel is the LAND colour and
/// the two arms differ over nearly the whole pane. With the constant lowered below every
/// reachable way rank it is the water colour and the two renders agree.
#[test]
fn a_tunnel_in_the_loaded_set_cannot_put_the_white_land_over_the_map() {
    if !gated("a_tunnel_in_the_loaded_set_cannot_put_the_white_land_over_the_map") {
        return;
    }
    let land_col = LAND_WHITE.to_array();

    let shot = |tunnel: bool, name: &str| {
        let (frames, img) = drive(land_lake_and_road(tunnel), 3);
        let path = format!("{}/scratch/cargo/gpu_lane_landmass_{name}.png", std::env::var("HOME").unwrap_or_default());
        if image::DynamicImage::ImageRgba8(img.clone()).save(&path).is_ok() {
            eprintln!("LOOK: saved {path} ({}×{})", img.width(), img.height());
        }
        (frames, img)
    };
    let (f_plain, plain) = shot(false, "surface_road");
    let (f_tunnel, tunnel) = shot(true, "tunnel");

    // ── NON-VACUITY ──────────────────────────────────────────────────────────
    assert!(
        f_plain > 0 && f_tunnel > 0,
        "both renders must come off the GPU LANE (frames {f_plain} / {f_tunnel}); a \
         CPU-painter frame never places the callback and so says nothing about the seam"
    );
    let centre_plain = px_at(&plain, 0.5, 0.5);
    let centre_tunnel = px_at(&tunnel, 0.5, 0.5);
    // The landmass really painted: a corner well outside the lake is the LAND colour in
    // the control. Without this the "centre is not land" claim below could pass on a
    // facet that never drew a landmass at all.
    // Bottom-right: outside the lake (roughly x 144..376, y 84..316 at this camera) and
    // clear of every corner an overlay claims — the lane flag takes TOP-LEFT and occupies
    // y = 10..25, which is exactly where a (0.05, 0.05) probe lands. Measured, not
    // reasoned: the first spelling of this probe read the chip's plate [19, 36, 25] and
    // failed against a land colour that was on screen the whole time.
    let corner = px_at(&plain, 0.95, 0.95);
    assert!(
        near(corner, land_col, 8),
        "the landmass must really be on screen in the control: the corner is {corner:?}, \
         the land colour is {land_col:?} — with no white land there is nothing to paint over"
    );
    // ... and the centre of the control is the WATER, not the land, so the two colours
    // are distinguishable at the probe pixel.
    assert!(
        !near(centre_plain, land_col, 8),
        "CONTROL BROKEN: with an ordinary surface road the pane centre is already the \
         land colour ({centre_plain:?} vs {land_col:?}) — the lake never painted and the \
         claim below is untestable"
    );

    // ── THE CPU ARM, measured in the same run ────────────────────────────────
    //
    // `korp-ui-cpu` (the same UI without the wgpu leg) cannot show this defect, and the
    // reason is structural rather than lucky: with no lane there is no SPLIT — `ui` does
    // `gpu_floor.map_or(feats.len(), ..)` (layer.rs:2565) — so the CPU list paints as one
    // uninterrupted `PaintKey`-sorted run and the landmass is first whatever its band.
    // Asserted here so "right in CPU, wrong in GPU" is a measurement in this file and not
    // a claim about a binary nobody ran.
    let cpu_arm = land_lake_and_road(true);
    let cpu_feats = cpu_arm.features_for_cpu_for_test(&[], false, false);
    let (cpu_floor, cpu_split) = cpu_arm.gpu_slot_split_for_test(&[], false, false, &cpu_feats);
    assert!(!cpu_feats.is_empty(), "the CPU arm must build features at all");
    assert_eq!(cpu_floor, None, "with no lane there is no rank floor to place a slot at");
    assert_eq!(
        cpu_split,
        cpu_feats.len(),
        "with no lane every CPU feature must paint in ONE run before the (absent) callback \
         — that is why the pure-CPU build never shows the white land"
    );

    let diff = plain.pixels().zip(tunnel.pixels()).filter(|(a, b)| a != b).count();
    let total = (plain.width() * plain.height()) as usize;
    let ok = !near(centre_tunnel, land_col, 8);
    eprintln!(
        "LOOK: centre surface-road {centre_plain:?}, centre tunnel {centre_tunnel:?}, \
         land {land_col:?}; {diff} of {total} pixels differ between the two arms"
    );
    nornir_testmatrix::functional_status(
        "facett-map",
        "landmass_stays_under_the_gpu_slot",
        ok,
        &format!(
            "centre: surface {centre_plain:?} / tunnel {centre_tunnel:?}, land {land_col:?}; \
             {diff}/{total} px differ; lane frames {f_plain}/{f_tunnel}"
        ),
    );
    assert!(
        ok,
        "ONE `tunnel=yes` way in the loaded set turned the pane centre into the WHITE \
         LAND ({centre_tunnel:?} vs land {land_col:?}); the same fixture with a surface \
         road shows {centre_plain:?}. {diff} of {total} pixels differ between two renders \
         whose only difference is one OSM tag. `paint_order::LANDMASS_BAND` must sit below \
         every rank `way_rank` can produce (-LAYER_CLAMP * LAYER_STEP), or a negative \
         `layer`/`tunnel` drags the GPU lane's slot under the land silhouette."
    );
}

/// **Two ROADS crossing: the major one must be on top** — lane MP, the owner 2026-09-05,
/// live on t14s (Radeon): *"roads are not painted in z-led order, it looks Turbo Pascal
/// 1993 … minor roads overpaint motorways/bridges."*
///
/// # Why this case was not covered
///
/// The five oracles above all cross a **fill** with a **line** (a lake and a road, a
/// tile boundary, the landmass). Every one of them passes on this box. **None of them
/// crosses two LINES of different class**, which is exactly what the owner reports, and it
/// exercises a different part of the machinery: both ways land in the same submission
/// stream and the same draw, so nothing about the fills/outlines/lines bucket order can
/// decide it. Only `PaintKey` order inside the stream can.
///
/// MEASURED before writing this, so the claim is about the renderer and not the ladder:
/// `way_rank` gives motorway **1240** and residential **1210** (secondary 1230, tertiary
/// 1220, service 1200, footway 1100), and `gpu/mod.rs` sorts `line_polys` by the full
/// `PaintKey`. So the ORDER is right and this asks whether the PIXELS agree.
///
/// The residential is drawn wider than the motorway on purpose. If it painted last it
/// would cover the motorway completely, and a width tie would let a correct result and a
/// broken one look the same on the crossing pixel.
#[test]
fn a_minor_road_does_not_paint_over_a_major_one() {
    if !gated("a_minor_road_does_not_paint_over_a_major_one") {
        return;
    }
    // Crossing at the pane centre: the motorway east-west, the residential north-south.
    let major = line(1, 18.040, 18.080, 59.330, &[("highway", "motorway")]);
    let mut minor = line(2, 18.060, 18.060, 59.330, &[("highway", "residential")]);
    minor.lonlat = vec![(18.060, 59.310), (18.060, 59.350)];
    minor.width = major.width * 3.0;
    assert!(
        facett_map::way_rank(&major) > facett_map::way_rank(&minor),
        "fixture: the motorway must OUTRANK the residential, or this proves nothing \
         ({} vs {})",
        facett_map::way_rank(&major),
        facett_map::way_rank(&minor)
    );
    let major_colour = major.stroke.to_array();
    let minor_colour = minor.stroke.to_array();
    assert!(
        !near(major_colour, minor_colour, 6),
        "fixture: the two classes must be visibly different colours, or the crossing \
         pixel cannot tell them apart ({major_colour:?} vs {minor_colour:?})"
    );

    let view = MapView::new(LatLon::new(59.33, 18.06), 13.0, (CANVAS.x, CANVAS.y));
    let facet = MapFacet::new("gpu-road-order", view, MapTheme::default())
        .with_layer(Box::new(WaysLayer::new("osm-ways", vec![minor, major])));
    let (frames, img) = drive(facet, 3);
    let path = format!("{}/scratch/cargo/gpu_lane_paint_order_road_over_road.png", std::env::var("HOME").unwrap_or_default());
    let _ = image::DynamicImage::ImageRgba8(img.clone()).save(&path);
    eprintln!("LOOK: saved {path}");
    assert!(frames > 0, "the GPU lane encoded ZERO frames — this is the CPU painter");

    let crossing = px_at(&img, 0.5, 0.5);
    let ok = near(crossing, major_colour, 40);
    nornir_testmatrix::functional_status(
        "facett-map",
        "gpu_lane_major_road_over_minor",
        ok,
        &format!(
            "{frames} lane frames; crossing {crossing:?}, motorway {major_colour:?}, \
             residential {minor_colour:?}"
        ),
    );
    assert!(
        ok,
        "the MOTORWAY must own the crossing pixel: got {crossing:?}, motorway is \
         {major_colour:?}, residential is {minor_colour:?}"
    );
}

/// **A minor road in a LATER tile must not paint over a motorway in an earlier one** —
/// lane MP, the case the suite did not cover.
///
/// # How this narrows the owner's report
///
/// `a_minor_road_does_not_paint_over_a_major_one` (above) puts both roads in ONE tile
/// and **passes**: inside a single submission stream `PaintKey` order decides, and it
/// decides correctly. `which_tile_a_feature_landed_in_cannot_decide_what_covers_what`
/// crosses a tile boundary and also passes — but with two **fills**.
///
/// **Nothing crossed a tile boundary with two LINES**, and that is where the machinery
/// differs: `gpu/mod.rs` draws lines with `for chunk in &renderer.line_chunks { … }`, one
/// `draw_indirect` per chunk, and the tiled path iterates **per tile**. `PaintKey` orders
/// the stream *inside* one submission and has no way to reach across two — the module
/// says so itself. So a minor road submitted in a later tile can land on top of a
/// motorway submitted in an earlier one, and no amount of correct ranking prevents it.
///
/// That is the shape of *"minor roads overpaint motorways"* on a real Stockholm camera
/// (many tiles) while every isolated fixture is green.
///
/// Run BOTH arrangements: which tile the motorway happens to fall in must not decide
/// what covers what. If only one order is tested, a broken renderer passes half the time.
#[test]
fn which_tile_a_road_landed_in_cannot_decide_which_road_is_on_top() {
    if !gated("which_tile_a_road_landed_in_cannot_decide_which_road_is_on_top") {
        return;
    }
    for swap in [false, true] {
        let mut major = line(1, 18.040, 18.080, 59.330, &[("highway", "motorway")]);
        let mut minor = line(2, 18.060, 18.060, 59.330, &[("highway", "residential")]);
        minor.lonlat = vec![(18.060, 59.310), (18.060, 59.350)];
        minor.width = major.width * 3.0;
        assert!(
            facett_map::way_rank(&major) > facett_map::way_rank(&minor),
            "fixture: the motorway must outrank the residential"
        );
        // The motorway is in the LATE tile in one arrangement and the EARLY one in the
        // other. Neither may change the answer.
        major.tile_key = if swap { 2 } else { 1 };
        minor.tile_key = if swap { 1 } else { 2 };
        let major_colour = major.stroke.to_array();
        let minor_colour = minor.stroke.to_array();

        let view = MapView::new(LatLon::new(59.33, 18.06), 13.0, (CANVAS.x, CANVAS.y));
        let facet = MapFacet::new("gpu-road-tile-order", view, MapTheme::default())
            .with_layer(Box::new(WaysLayer::new("osm-ways", vec![minor, major])));
        let (frames, img) = drive(facet, 3);
        let path = format!(
            "{}/scratch/cargo/gpu_lane_road_tiles_{}.png",
            std::env::var("HOME").unwrap_or_default(),
            if swap { "major_late" } else { "major_early" }
        );
        let _ = image::DynamicImage::ImageRgba8(img.clone()).save(&path);
        eprintln!("LOOK: saved {path}");
        assert!(frames > 0, "swap={swap}: the GPU lane encoded ZERO frames");

        let crossing = px_at(&img, 0.5, 0.5);
        let ok = near(crossing, major_colour, 40);
        nornir_testmatrix::functional_status(
            "facett-map",
            "gpu_lane_road_over_road_across_tiles",
            ok,
            &format!(
                "swap={swap}: crossing {crossing:?}, motorway {major_colour:?}, \
                 residential {minor_colour:?}, {frames} frames"
            ),
        );
        assert!(
            ok,
            "swap={swap}: the MOTORWAY must own the crossing pixel whichever tile it \
             landed in — got {crossing:?}, motorway {major_colour:?}, residential \
             {minor_colour:?}"
        );
    }
}

// ─────────────────────────────────────────────────────────────────────────────
// LANE GG #1 — the fill ring OUTLINE sits at z = 0, the closest plane there is
// ─────────────────────────────────────────────────────────────────────────────

/// **A ROAD CROSSING A LANDCOVER RING'S EDGE MUST STILL BE THE ROAD.**
///
/// # The defect, and that it is mine
///
/// `build_gpu_fill_geometry` pushes every ring outline through
/// [`line::LineExpander::push_pts`], whose depth argument is **`0.0`** — and `0.0` is not
/// a neutral default, it is the CLOSEST plane the clip volume has. Against a depth buffer
/// cleared to `1.0` under `Less`, the 0.5 px outline of every lake, every park and every
/// building wins **every pixel it touches, against everything drawn afterwards,
/// regardless of rank**.
///
/// That is a regression of the MP-a offscreen depth pass I landed this morning. Before
/// it there was no depth buffer and the three hard buckets decided the stack, so an
/// outline at `0.0` cost nothing; the moment depth started deciding, the same constant
/// became the highest-priority ink on the map. The audit found it (lane GG #1) and it is
/// the mechanism behind what the owner actually sees: **OSM landuse rings share their
/// geometry with the road that bounds them**, so this edge does not sit near the road, it
/// sits ON the road's axis — a forest boundary drawn along a lane of tarmac erases it.
///
/// # The fixture
///
/// A `landuse=forest` ring whose RIGHT edge runs exactly through the camera centre, and
/// a motorway straight across it at the centre latitude. So the pane centre is the one
/// pixel where road and ring outline both land, and `(0.75, 0.5)` is the same road over
/// bare paper — the control that keeps this from passing by the road simply vanishing.
///
/// The red arm is the constant itself: restore `0.0` and the crossing pixel becomes the
/// outline.
#[test]
fn a_road_crossing_a_landcover_ring_edge_is_still_the_road() {
    if !gated("a_road_crossing_a_landcover_ring_edge_is_still_the_road") {
        return;
    }
    // Right edge at lon 18.060 = the camera's centre longitude.
    let forest = ring(1, 18.040, 59.330, 0.020, &[("landuse", "forest")]);
    let road = line(2, 18.030, 18.090, 59.330, &[("highway", "motorway")]);
    assert!(forest.fill.is_some(), "fixture: the forest must be a FILL, or it has no outline");
    assert_eq!(road.style.kind, WayKind::Road, "fixture: the line really classifies as a road");
    assert!(
        facett_map::way_rank(&road) > facett_map::way_rank(&forest),
        "fixture: the motorway must outrank the forest, or there is no ordering claim here \
         at all — road {} vs forest {}",
        facett_map::way_rank(&road),
        facett_map::way_rank(&forest),
    );
    let road_colour = road.stroke.to_array();
    let outline_colour = forest.style.color.to_array();

    let view = MapView::new(LatLon::new(59.33, 18.06), 13.0, (CANVAS.x, CANVAS.y));
    // The forest is fed LAST, so "whoever was read last wins" would hand it the pixel.
    let facet = MapFacet::new("gpu-ring-edge", view, MapTheme::default())
        .with_layer(Box::new(WaysLayer::new("osm-ways", vec![road, forest])));
    let (frames, img) = drive(facet, 3);
    let path = format!("{}/scratch/cargo/gpu_lane_ring_edge.png", std::env::var("HOME").unwrap_or_default());
    let _ = image::DynamicImage::ImageRgba8(img.clone()).save(&path);
    eprintln!("LOOK: saved {path}");
    assert!(frames > 0, "the GPU lane encoded ZERO frames — this says nothing about the device");

    let crossing = px_at(&img, 0.5, 0.5);
    let road_clear = px_at(&img, 0.75, 0.5);
    let forest_clear = px_at(&img, 0.30, 0.25);
    let edge_clear = px_at(&img, 0.5, 0.25);

    // NON-VACUITY, all three of them. Without these the claim below is satisfiable by a
    // render that drew nothing at all.
    assert!(
        near(road_clear, road_colour, 40),
        "the road must actually be on screen away from the ring: {road_clear:?} vs \
         {road_colour:?}"
    );
    assert!(
        !near(forest_clear, road_clear, 6),
        "the forest fill {forest_clear:?} is indistinguishable from the road {road_clear:?} \
         — nothing here can be told apart, so nothing is proved"
    );
    assert!(
        !near(edge_clear, forest_clear, 6),
        "the ring OUTLINE must be visible where no road crosses it ({edge_clear:?} vs the \
         fill's own {forest_clear:?}) — if the outline is not drawn at all, the defect this \
         file is about cannot be reproduced and a green here would be a lie"
    );

    let ok = near(crossing, road_clear, 24);
    nornir_testmatrix::functional_status(
        "facett-map",
        "gpu_lane_road_over_ring_edge",
        ok,
        &format!(
            "crossing {crossing:?}, road {road_clear:?}, ring outline {edge_clear:?}, \
             forest fill {forest_clear:?}, {frames} frames"
        ),
    );
    assert!(
        ok,
        "the ROAD must own the pixel where it crosses the forest ring's edge — got \
         {crossing:?}, the same road over paper is {road_clear:?}, the ring outline alone \
         is {edge_clear:?}. An outline pushed at depth 0.0 is the closest plane in the clip \
         volume and beats every rank drawn after it, which is exactly what the owner sees as \
         'roads overwritten by landcover'. Outline colour as declared: {outline_colour:?}"
    );
}

// ─────────────────────────────────────────────────────────────────────────────
// LANE GG #2 — the anti-aliasing fringe writes depth
// ─────────────────────────────────────────────────────────────────────────────

/// A road built for this claim: `major` is a DIAGONAL motorway through the centre in the
/// EARLY tile, `minor` a residential spanning the whole pane horizontally in the LATE one.
///
/// Diagonal on purpose. An axis-aligned road at this camera lands its edges on whole
/// pixels, so no fragment is ever partially covered and the halo this claim is about never
/// rasterises at all — measured first with a vertical motorway, and every pixel was either
/// full ink or pure paper while the test passed proving nothing. A diagonal edge crosses
/// pixel centres at every fraction.
fn fringe_facet(with_major: bool, with_minor: bool) -> MapFacet {
    let mut ways = Vec::new();
    if with_major {
        let mut major = line(1, 18.040, 18.080, 59.330, &[("highway", "motorway")]);
        major.lonlat = vec![(18.030, 59.300), (18.090, 59.360)];
        major.tile_key = 1;
        ways.push(major);
    }
    if with_minor {
        let mut minor = line(2, 17.990, 18.130, 59.330, &[("highway", "residential")]);
        minor.tile_key = 2;
        ways.push(minor);
    }
    let view = MapView::new(LatLon::new(59.33, 18.06), 13.0, (CANVAS.x, CANVAS.y));
    MapFacet::new("gpu-aa-fringe-depth", view, MapTheme::default())
        .with_layer(Box::new(WaysLayer::new("osm-ways", ways)))
}

/// **AN INVISIBLE FRAGMENT MUST NOT CLAIM A PIXEL.**
///
/// # The defect
///
/// `line.wgsl`'s quad is rasterised one `feather_px` WIDER than the road on each side, so
/// the AA ramp has fragments to land on out to zero coverage — correct, and the reason
/// the drawn width comes out exact. `fs_main` faded alpha to `0.0` across that band and
/// did **not** discard, so every road carried a ~1 px halo of invisible ink that still
/// passed and wrote depth. Any lower-ranked way drawn afterwards — a later tile, a later
/// chunk — was depth-rejected there, and a rejected fragment is not blended away: what
/// remains is the bare pane background, a paper-coloured slit alongside every road.
///
/// It cost nothing until MP-a's depth pass turned depth on. Nobody was reading what this
/// fringe wrote.
///
/// # What the claim is, and what it deliberately is NOT
///
/// It is **not** "no pixel of the road beneath shows through to paper". That claim is
/// false and would be wrong to assert: with depth testing on, the alpha-blended edge of a
/// *higher*-ranked way rejects the way beneath, so a genuinely half-covered edge pixel
/// shows motorway-over-paper rather than motorway-over-residential. The first version of
/// this test asserted exactly that and stayed red after the fix, for a reason that is not
/// this defect.
///
/// The claim is exact instead: **the set of pixels the motorway takes from the road
/// beneath must be the same set it visibly paints.** Measured with three renders of the
/// same camera — both roads, the residential alone, the motorway alone — so both sets are
/// differences against a control rather than colour guesses with a tolerance.
///
/// # HONEST STATUS: this gate was GREEN before the discard as well
///
/// Driven both ways on oden / RTX 4090, 2026-09-05: **taken = 7, painted = 7, the same
/// seven pixels, with the `discard` in and with it neutralised.** So the discard changes
/// no pixel here, and the audit's stated mechanism for GG #2 — *"a 1 px transparent halo
/// stamps depth"* — **does not survive measurement**. It cannot: the quad extends one
/// feather past the road and `coverage_inside` ramps 0.5 → 0 across that band, so a
/// fragment under half an alpha step needs its centre within ~0.002 px of the outer edge.
/// Such a fragment essentially never rasterises. The `discard` is kept because an
/// invisible fragment reserving a pixel is wrong by construction and costs nothing, not
/// because it was measured to fix anything.
///
/// **What DOES cut the road beneath, measured on the same frames:** at x = 256 the
/// motorway's casing covers ~35 % of the pixel. `both` reads `[28, 28, 34]` — casing over
/// PAPER. Blended correctly over the residential it would be `[38, 41, 50]`. The
/// residential was not blended with, it was **discarded by the depth test**, and that is
/// the "1 px casing cut" symptom: not a transparent halo, but every partially covered
/// fragment of a higher-ranked way erasing the lower one instead of compositing with it.
/// It only bites when the lower-ranked way is drawn *afterwards* — across tiles — because
/// inside one tile `PaintKey` already orders them correctly and the AA composites fine.
///
/// Closing that needs the AA edge to stop writing depth, which is a real change with a
/// real cost (4× MSAA on the offscreen target, audit #32; or a two-pass line draw with
/// the edge depth-write-off). Recorded in korp-v2 rather than picked here.
///
/// So what this gate is worth: it pins the invariant *"a way takes only what it paints"*
/// against a future regression of the GG #1 kind, and it carries the numbers above so the
/// next reader does not re-derive them.
#[test]
fn the_aa_fringe_of_a_road_does_not_cut_a_slit_in_the_road_beneath_it() {
    if !gated("the_aa_fringe_of_a_road_does_not_cut_a_slit_in_the_road_beneath_it") {
        return;
    }
    let (frames, both) = drive(fringe_facet(true, true), 3);
    let (_, minor_only) = drive(fringe_facet(false, true), 3);
    let (_, major_only) = drive(fringe_facet(true, false), 3);
    let path = format!("{}/scratch/cargo/gpu_lane_aa_fringe.png", std::env::var("HOME").unwrap_or_default());
    let _ = image::DynamicImage::ImageRgba8(both.clone()).save(&path);
    eprintln!("LOOK: saved {path}");
    assert!(frames > 0, "the GPU lane encoded ZERO frames — this says nothing about the device");

    let y = both.height() / 2;
    let px = |img: &image::RgbaImage, x: u32| img.get_pixel(x, y).0;
    // The paper this camera actually renders, off the same row, well clear of the
    // diagonal — read from the residential-only frame at its far edge would be the road,
    // so take it from a row neither way touches.
    let paper = px_at(&major_only, 0.05, 0.90);

    // TAKEN: where adding the motorway changed the residential.
    let taken: Vec<u32> =
        (0..both.width()).filter(|&x| !near(px(&both, x), px(&minor_only, x), 2)).collect();
    // PAINTED: where the motorway alone changed the paper. Same row, same camera, same
    // diagonal — so the two counts are directly comparable.
    //
    // `alpha == 255` is the pane. The harness canvas is wider than the map pane and its
    // margin is genuinely unpainted (`[0, 0, 0, 0]`), which is not paper and would
    // otherwise count as sixteen pixels of "the motorway painted here" at the two ends of
    // every row — enough on its own to make this comparison pass whatever the shader did.
    // It did: measured 23 painted against 7 taken, and 16 of the 23 were margin.
    //
    // **The tolerance is 1, and it was 2 until lane GG #2's cure landed.** Not a loosening
    // to make something pass — a recalibration onto the noise floor, and the two numbers
    // it sits between were measured on this exact row on oden / RTX 4090, 2026-09-05:
    //
    // * The paper this camera renders dithers by exactly ONE level (`[31, 31, 37]` beside
    //   `[31, 31, 38]`, nine such pixels across the row). At tolerance 0 every one of them
    //   is a phantom "the motorway painted here" and the set is meaningless.
    // * At tolerance 2 the set drops a pixel the motorway genuinely paints. GG #2's blit
    //   fix (`gpu/mod.rs`: the offscreen is premultiplied, so the blit must be too) removed
    //   an `α²` that had been over-darkening every partial pixel, and the motorway's
    //   faintest edge at x = 256 went `[28, 28, 34]` → `[29, 29, 36]` against paper
    //   `[31, 31, 38]` — from 4 levels of contrast to 2, i.e. exactly onto the boundary.
    //   It still paints there, and it still takes there (12 levels off the residential),
    //   so `taken > painted` was the instrument losing resolution, not a way claiming a
    //   pixel it never painted.
    //
    // 1 is the floor of the paper's own noise and the only value that separates the two.
    let painted: Vec<u32> = (0..major_only.width())
        .filter(|&x| px(&major_only, x)[3] == 255 && !near(px(&major_only, x), paper, 1))
        .collect();

    // NON-VACUITY. Both sets must be real, and the residential must actually be under the
    // motorway on this row — otherwise "taken == painted" is 0 == 0.
    assert!(
        !painted.is_empty(),
        "the motorway paints nothing on this row: nothing is being compared"
    );
    assert!(
        !taken.is_empty(),
        "the motorway takes nothing from the residential: the two are not crossing"
    );
    assert!(
        !near(px(&minor_only, 20), paper, 12),
        "the residential did not paint on this row ({:?} against paper {paper:?})",
        px(&minor_only, 20)
    );

    let ok = taken.len() <= painted.len();
    nornir_testmatrix::functional_status(
        "facett-map",
        "gpu_lane_aa_fringe_writes_no_depth",
        ok,
        &format!(
            "took {} px from the road beneath, painted {} px; {frames} frames",
            taken.len(),
            painted.len()
        ),
    );
    assert!(
        ok,
        "the motorway took {} pixels from the residential but visibly paints only {} — \
         {} of them it never painted at all. That is its zero-coverage AA fringe writing \
         depth: an invisible fragment reserving a pixel, and the road beneath discarded \
         rather than blended, leaving bare paper. taken {taken:?}, painted {painted:?}",
        taken.len(),
        painted.len(),
        taken.len() - painted.len()
    );
}

/// **A PARTIALLY COVERED EDGE MUST COMPOSITE WITH WHAT IS BENEATH IT, NOT ERASE IT.**
///
/// # The claim its sibling above deliberately does not make
///
/// `the_aa_fringe_of_a_road_does_not_cut_a_slit_in_the_road_beneath_it` compares two
/// pixel SETS — *taken* against *painted* — and is green whether or not the compositing
/// is right, because a pixel the motorway erases and a pixel the motorway blends are
/// both "taken". Its own header says so and files the stronger claim as unfixed.
///
/// This is that stronger claim, and it is about COLOUR: at a pixel the motorway only
/// PARTIALLY covers, what must be on screen is the motorway composited over the
/// **residential**, not over the paper.
///
/// # The mechanism, and why it is not the AA halo
///
/// `line.wgsl` rasterises the quad one `feather_px` wider than the road and takes
/// fragment alpha from `coverage_inside`. At a pixel straddling the road's true edge the
/// quad covers the pixel COMPLETELY — the extra feather of geometry is still ahead of it
/// — so the fragment is not a fringe fragment at all: it is an ordinary, fully
/// rasterised fragment carrying alpha ≈ 0.35. Depth writing is on and unconditional, so
/// it stamps the whole pixel's depth at the motorway's Z. The residential, drawn
/// afterwards because it lives in another tile, is then depth-REJECTED there — and a
/// rejected fragment is not blended away, it never runs. What survives is the motorway
/// over bare paper.
///
/// So the fragment had to choose between compositing and owning the pixel, and depth
/// made it own the pixel. That is a coverage problem, and the fix is to give the coverage
/// somewhere to live other than the alpha: the offscreen colour+depth pair goes 4×
/// (`gpu::MAP_MSAA_SAMPLES`) so depth is per SAMPLE, and `line.wgsl` hands its own
/// `coverage_inside` to the rasteriser as `@builtin(sample_mask)`, so a ⅜-covered edge
/// claims one or two samples of four and leaves the rest for whatever is underneath.
///
/// **The 4× target alone does not do it**, and the failed attempt is worth carrying: with
/// the target and every pipeline at 4× and the shader untouched, this test's frames came
/// back byte-for-byte identical to the single-sample ones, for exactly the reason in the
/// paragraph above — the rasteriser sees full coverage at these pixels because the quad is
/// a feather wider than the road. Alpha-to-coverage, the one-line way to hand the alpha
/// back, was tried too and is worse: it multiplies the coverage in twice and dithers a
/// road's fill against its own casing. `gpu::map_msaa_state` carries those measurements.
///
/// # The probe
///
/// Three renders of one camera (both roads / residential alone / motorway alone) on the
/// pane's middle row, which runs along the residential's own centre line. The motorway's
/// footprint on that row is one contiguous run of pixels, and its two END pixels are
/// partial **by geometry**: the road's edge falls inside them, or the pixel beyond would
/// have been painted too. No colour threshold picks them and no measurement of this
/// defect chose them.
///
/// What is measured at each is
///
/// ```text
///   survival = |both − major_only| / |residential − paper|
/// ```
///
/// — how much the residential still MOVES the pixel, as a fraction of its own contrast
/// against paper. It is deliberately read as a magnitude and not as an exact `1 − a`:
/// what lies beneath is another cased road, not a flat colour, so which of the four
/// samples each way claims decides whether the residential's fill or its darker casing
/// reaches the resolve, and `both` can legitimately land on either side of `major_only`.
/// The one thing that needs no model is the floor: under the defect `both` and
/// `major_only` are the same pixel to the last bit, and survival is exactly `0.000` — not
/// at the ends, but at **every** pixel of the footprint.
///
/// MEASURED on oden / RTX 4090, 2026-09-05, at 4× with the sample mask in
/// (`gpu/line.wgsl`): `0.375` at the leading edge and `0.281` at the trailing one, against
/// `0.000` and `0.000` before it. The bar below sits at less than half the smaller.
#[test]
fn a_partially_covered_road_edge_composites_with_the_road_beneath_it() {
    if !gated("a_partially_covered_road_edge_composites_with_the_road_beneath_it") {
        return;
    }
    let (frames, both) = drive(fringe_facet(true, true), 3);
    let (_, minor_only) = drive(fringe_facet(false, true), 3);
    let (_, major_only) = drive(fringe_facet(true, false), 3);
    assert!(frames > 0, "the GPU lane encoded ZERO frames — this says nothing about the device");

    let y = both.height() / 2;
    let px = |img: &image::RgbaImage, x: u32| img.get_pixel(x, y).0;
    let paper = px_at(&major_only, 0.05, 0.90);
    // Largest per-channel difference — the same three channels `near` compares, read as a
    // magnitude instead of a verdict.
    let spread = |a: [u8; 4], b: [u8; 4]| {
        (0..3).map(|i| (i32::from(a[i]) - i32::from(b[i])).abs()).max().unwrap_or(0)
    };

    // **The window is `taken` — where adding the motorway changed the residential — and
    // NOT the sibling test's `painted`.**
    //
    // That is a correction worth stating, because `painted` is the obvious choice and it
    // is quietly the wrong one: it selects pixels by how strongly the motorway's own ink
    // reads against paper, which is inversely related to how much of the pixel is left for
    // the road beneath. MEASURED here: at x = 256 the motorway is faint enough
    // (`[29, 29, 36]` against paper `[31, 31, 38]`) to fall inside `near`'s tolerance and
    // be dropped from `painted` — and that pixel is the one carrying the largest surviving
    // residential of the whole crossing. A probe that discards its best evidence for being
    // faint is measuring the wrong thing.
    //
    // `taken` is the population the claim is actually about, and it is what the sibling
    // test already counts, so the two files are looking at one set.
    let taken: Vec<u32> =
        (0..both.width()).filter(|&x| !near(px(&both, x), px(&minor_only, x), 2)).collect();

    // NON-VACUITY, before any claim about compositing.
    assert!(
        taken.len() >= 3,
        "the motorway changes only {} px of the residential on this row — there is no \
         interior for its two end pixels to be the partial ones OF. taken {taken:?}",
        taken.len()
    );
    assert!(
        taken.last().expect("non-empty") - taken[0] + 1 == taken.len() as u32,
        "the crossing on this row is not one contiguous run ({taken:?}) — the 'the end \
         pixels are the partial ones' argument is a geometric one and needs it to be"
    );
    for x in [taken[0], *taken.last().expect("non-empty")] {
        assert!(
            !near(px(&minor_only, x), paper, 12),
            "the residential does not paint at x={x} ({:?} against paper {paper:?}) — there \
             would be nothing underneath for the motorway to composite with",
            px(&minor_only, x)
        );
    }

    let edges = [taken[0], *taken.last().expect("non-empty")];
    let survival = |x: u32| {
        let available = spread(px(&minor_only, x), paper);
        let survived = spread(px(&both, x), px(&major_only, x));
        (survived, available, f64::from(survived) / f64::from(available.max(1)))
    };

    eprintln!(
        "[edge] row {y}, paper {paper:?}, footprint {taken:?}\n{}",
        taken
            .iter()
            .map(|&x| {
                let (s, a, k) = survival(x);
                format!(
                    "  x={x:3}  both {:?}  major {:?}  minor {:?}  survived {s:3}/{a:3} = {k:.3}",
                    px(&both, x),
                    px(&major_only, x),
                    px(&minor_only, x)
                )
            })
            .collect::<Vec<_>>()
            .join("\n")
    );

    // The bar: less than half the smaller of the two measured survivals (0.281), and far
    // above the only number the defect can produce, which is an exact 0.000. It is set as
    // a fraction rather than in levels so it does not silently weaken if the fixture's
    // palette or camera changes the residential's contrast against paper.
    const FLOOR: f64 = 0.15;
    let worst = edges
        .iter()
        .map(|&x| (x, survival(x)))
        .min_by(|a, b| a.1.2.total_cmp(&b.1.2))
        .expect("two edges");
    let ok = worst.1.2 > FLOOR;
    nornir_testmatrix::functional_status(
        "facett-map",
        "gpu_lane_partial_edge_composites",
        ok,
        &format!(
            "worst edge x={} survival {:.3} (survived {} of {} available); {frames} frames",
            worst.0, worst.1.2, worst.1.0, worst.1.1
        ),
    );
    assert!(
        ok,
        "at x={} the motorway only PARTIALLY covers the pixel, and the residential beneath \
         survived {:.3} of its own contrast against paper (needed > {FLOOR}). both {:?}, \
         motorway-over-paper {:?}, residential {:?}, paper {paper:?}.\n\
         A survival of 0 means the two frames are the same pixel: the residential was not \
         blended with, it was DISCARDED by the depth test, because a partially covered \
         fragment writes depth for the WHOLE pixel. What fixes that is `line.wgsl` handing \
         its coverage to the rasteriser as a `@builtin(sample_mask)` over the 4× offscreen \
         pair (`gpu::MAP_MSAA_SAMPLES`), so depth is claimed per SAMPLE. Note that the 4× \
         target ALONE does not fix it and was measured not to: the quad is rasterised one \
         feather wider than the road, so the rasteriser sees full coverage at exactly the \
         pixels this test probes.",
        worst.0,
        worst.1.2,
        px(&both, worst.0),
        px(&major_only, worst.0),
        px(&minor_only, worst.0),
    );
}

// ─────────────────────────────────────────────────────────────────────────────
// LANE GG #3 — `Less` inverts the order INSIDE a band
// ─────────────────────────────────────────────────────────────────────────────

/// **THE SMALL POLYGON INSIDE THE BIG ONE MUST BE ON TOP.**
///
/// # The defect
///
/// The tile lane pins a whole band RUN of fills to one depth through `set_viewport`
/// (`gpu/mod.rs:2221`) — correct, because `way_draw_band` for a fill IS its class, so a
/// run has nothing left to separate. Inside the run the stream is still ordered by
/// `PaintKey`, whose second term is **area, largest first**: a courtyard inside a block,
/// a pond inside a park, an island in a lake — the smaller polygon is emitted later
/// precisely so it lands on top.
///
/// `Less` throws that away. Every fragment in the run carries the same depth, the first
/// one writes it, and every later one is **rejected for being equal**. So the order
/// inside a band is not merely lost, it is exactly **inverted**: the biggest polygon
/// wins every overlap it is part of.
///
/// Like GG #1 and #2 this is a consequence of MP-a turning depth on — with no depth
/// buffer the ordered stream decided the stack by itself, which is what `PaintKey` was
/// built for.
///
/// # The fixture
///
/// Two buildings, one inside the other, so they share a band by construction — same class
/// and same `layer`, which is the only way to be in one run. They therefore also share a
/// palette colour, so the inner one's `fill` is overridden to something distinguishable;
/// the override touches neither `style.kind` nor the `layer` tag, so the band, the rank
/// and the emitted order are exactly what the palette would have produced.
#[test]
fn a_small_fill_inside_a_big_one_of_the_same_band_is_on_top() {
    if !gated("a_small_fill_inside_a_big_one_of_the_same_band_is_on_top") {
        return;
    }
    let big = ring(1, 18.060, 59.330, 0.020, &[("building", "yes")]);
    let mut small = ring(2, 18.060, 59.330, 0.004, &[("building", "yes")]);
    assert_eq!(
        facett_map::paint_order::way_draw_band(&big),
        facett_map::paint_order::way_draw_band(&small),
        "fixture: the two must share a BAND, or they are pinned to different depths and \
         this claim is about nothing"
    );
    // Distinguishable ink. Same class, same palette, so without this both rings are the
    // identical colour and any answer looks like the right one — the identity trap.
    let inner = egui::Color32::from_rgb(255, 0, 128);
    small.fill = Some(inner);
    small.style.color = inner;
    let outer = big.fill.expect("fixture: a building is a fill");
    assert!(
        !near(inner.to_array(), outer.to_array(), 24),
        "fixture: the two fills must be told apart: {:?} vs {:?}",
        inner.to_array(),
        outer.to_array()
    );
    assert!(
        facett_map::paint_order::way_paint_key(&small)
            > facett_map::paint_order::way_paint_key(&big),
        "fixture: the SMALLER ring must sort LATER — that is the area tie-break this claim \
         is about, and if it does not hold the declared order already says the big one wins"
    );

    let view = MapView::new(LatLon::new(59.33, 18.06), 13.0, (CANVAS.x, CANVAS.y));
    // Fed big-last, so "whoever was read last wins" would also hand it to the big one.
    let facet = MapFacet::new("gpu-band-inner", view, MapTheme::default())
        .with_layer(Box::new(WaysLayer::new("osm-ways", vec![small.clone(), big])));
    let (frames, img) = drive(facet, 3);
    let path = format!("{}/scratch/cargo/gpu_lane_band_inner.png", std::env::var("HOME").unwrap_or_default());
    let _ = image::DynamicImage::ImageRgba8(img.clone()).save(&path);
    eprintln!("LOOK: saved {path}");
    assert!(frames > 0, "the GPU lane encoded ZERO frames");

    let centre = px_at(&img, 0.5, 0.5);
    // NON-VACUITY: the big ring really painted, somewhere the small one is not.
    let outside = px_at(&img, 0.5, 0.36);
    assert!(
        near(outside, outer.to_array(), 24),
        "the big building did not paint outside the small one ({outside:?} vs \
         {:?}) — nothing is being covered",
        outer.to_array()
    );

    let ok = near(centre, inner.to_array(), 24);
    nornir_testmatrix::functional_status(
        "facett-map",
        "gpu_lane_inner_fill_of_a_band_wins",
        ok,
        &format!(
            "centre {centre:?}, inner {:?}, outer {outer:?}, {frames} frames",
            inner.to_array()
        ),
    );
    assert!(
        ok,
        "the SMALL building inside the big one must own the centre: got {centre:?}, inner \
         is {:?}, the big one is {:?}. A whole band run of fills is pinned to ONE depth, so \
         `Less` rejects every later fragment for being EQUAL — which inverts the area \
         tie-break and hands every containment to the bigger polygon: pond in park, \
         courtyard in block, island in lake.",
        inner.to_array(),
        outer.to_array()
    );
}

// ─────────────────────────────────────────────────────────────────────────────
// LANE GG #28 — a railway is a class of the road network, not overlay ink
// ─────────────────────────────────────────────────────────────────────────────

/// One `railway=rail` north-south through the pane centre, one road east-west across it,
/// on the device lane. Returns `(lane frames, crossing pixel, road ink, rail ink)`.
///
/// The LOSER of each pair is drawn three times wider, exactly as
/// `a_minor_road_does_not_paint_over_a_major_one` does: if the order were wrong, the wide
/// loser would own the crossing unambiguously rather than through an AA blend.
fn rail_crossing(road_tags: &[(&str, &str)], rail_wins: bool, shot: &str) -> (u64, [u8; 4], [u8; 4], [u8; 4]) {
    let mut road = line(1, 18.040, 18.080, 59.330, road_tags);
    let mut rail = line(2, 18.060, 18.060, 59.330, &[("railway", "rail")]);
    rail.lonlat = vec![(18.060, 59.310), (18.060, 59.350)];
    assert_eq!(road.style.kind, WayKind::Road, "fixture: the road really classifies as a road");
    assert_ne!(rail.style.kind, WayKind::Other, "fixture: a railway must classify as its own kind, not `Other`");
    assert!(rail.fill.is_none() && road.fill.is_none(), "fixture: two LINES");
    if rail_wins {
        assert!(
            facett_map::way_rank(&rail) > facett_map::way_rank(&road),
            "fixture: the rail ({}) must OUTRANK this road ({}), or this proves nothing",
            facett_map::way_rank(&rail),
            facett_map::way_rank(&road)
        );
        road.width = rail.width * 3.0;
    } else {
        assert!(
            facett_map::way_rank(&rail) < facett_map::way_rank(&road),
            "fixture: this road ({}) must OUTRANK the rail ({}), or this proves nothing",
            facett_map::way_rank(&road),
            facett_map::way_rank(&rail)
        );
        rail.width = road.width * 3.0;
    }
    let road_ink = road.stroke.to_array();
    let rail_ink = rail.stroke.to_array();
    assert!(
        !near(road_ink, rail_ink, 6),
        "fixture: the road and the rail must be visibly different inks, or the crossing \
         pixel cannot tell them apart ({road_ink:?} vs {rail_ink:?})"
    );
    let view = MapView::new(LatLon::new(59.33, 18.06), 13.0, (CANVAS.x, CANVAS.y));
    // The winner is fed FIRST, so "whoever was read last wins" hands the crossing to the
    // loser in both arrangements.
    let ways = if rail_wins { vec![rail, road] } else { vec![road, rail] };
    let facet = MapFacet::new("gpu-rail-order", view, MapTheme::default())
        .with_layer(Box::new(WaysLayer::new("osm-ways", ways)));
    let (frames, img) = drive(facet, 3);
    let path = format!("{}/scratch/cargo/gpu_lane_paint_order_{shot}.png", std::env::var("HOME").unwrap_or_default());
    let _ = image::DynamicImage::ImageRgba8(img.clone()).save(&path);
    eprintln!("LOOK: saved {path}");
    assert!(frames > 0, "{shot}: the GPU lane encoded ZERO frames — this is the CPU painter");
    (frames, px_at(&img, 0.5, 0.5), road_ink, rail_ink)
}

/// **A RAILWAY RUNS UNDER THE MOTORWAY AND OVER THE RESIDENTIAL STREET, on the lane that
/// ships** — GG audit row 28, *"rail has no kind"*.
///
/// # The defect
///
/// `WayKind` had no `Rail` arm. `classify_with`'s `railway` arm (`railtier-1`) resolved
/// the ink and the tier and then stamped `WayKind::Other`, whose line offset is
/// `LINE_BAND + 400` — the OVERLAY slot, above every road. On korp's
/// "sweden · roads+rails" every level crossing and every viaduct-less junction painted
/// the rail over the E4; the Garmin look, and osm2pgsql's own `z_order`, put it between
/// the primary and the secondary network.
///
/// Two renders, because one crossing proves only half of "between": the rail must LOSE
/// the pixel to a motorway and WIN it from a residential street.
///
/// RED before `WayKind::Rail` (t14s, Radeon 890M): the fixture's own rank precondition
/// fires — `railway=rail` ranked 1440 against the motorway's 1240.
///
/// # A rail has TWO inks along its core since HO-T3
///
/// The dash lane cuts a rail's core along its arc length, so a pixel on the centreline is
/// the core tone in a lit phase, the rail's own CASING in a gap, or — where the pixel
/// straddles a dash END, which `dash_coverage` feathers exactly like a side — a blend of
/// the two. The pane centre of this fixture is such an edge (MEASURED t14s / 890M,
/// 2026-09-06: crossing `[95, 96, 100]` = casing `[14, 14, 17]` + 0.62 × (core
/// `[145, 147, 152]` − casing) on all three channels, against the street's
/// `[85, 98, 117]`). Every point on that casing→core segment is the rail owning the
/// pixel; what must never be there is the street. So the second half accepts any rail ink
/// ([`on_blend`]) and refuses the street's explicitly, and checks first that the street
/// itself is nowhere on that segment, or the refusal would prove nothing.
#[test]
fn a_railway_runs_under_the_motorway_and_over_the_residential_street() {
    if !gated("a_railway_runs_under_the_motorway_and_over_the_residential_street") {
        return;
    }
    let (frames, crossing, motorway, rail) = rail_crossing(&[("highway", "motorway")], false, "rail_under_motorway");
    let ok = near(crossing, motorway, 40);
    nornir_testmatrix::functional_status(
        "facett-map",
        "gpu_lane_rail_under_motorway",
        ok,
        &format!("{frames} lane frames; crossing {crossing:?}, motorway {motorway:?}, rail {rail:?}"),
    );
    assert!(
        ok,
        "the MOTORWAY must own the crossing pixel over a railway: got {crossing:?}, \
         motorway is {motorway:?}, rail is {rail:?}"
    );

    let (frames, crossing, street, rail) = rail_crossing(&[("highway", "residential")], true, "rail_over_residential");
    // The rail's casing — its ink in a dash gap (HO-T3). `rail_crossing` paints with
    // `MapTheme::default()`, so this is the casing that frame actually carries.
    let casing = MapTheme::default().palette.casing.to_array();
    assert!(
        !on_blend(street, casing, rail, 8),
        "fixture: the residential ink {street:?} lies within 8 of the rail's casing→core \
         segment {casing:?}→{rail:?} — a rail pixel could not be told from the street, and \
         this proves nothing"
    );
    let ok = on_blend(crossing, casing, rail, 8) && !near(crossing, street, 8);
    nornir_testmatrix::functional_status(
        "facett-map",
        "gpu_lane_rail_over_residential",
        ok,
        &format!("{frames} lane frames; crossing {crossing:?}, residential {street:?}, rail {rail:?} / casing {casing:?}"),
    );
    assert!(
        ok,
        "the RAILWAY must own the crossing pixel over a residential street — its core \
         {rail:?} in a lit dash, its casing {casing:?} in a gap, or a blend of the two on a \
         dash edge: got {crossing:?}, residential is {street:?}"
    );
}

// ── MERGED CASINGS PER CLASS (lane GG #26) ──────────────────────────────────

/// A `highway=…` line from `(lon0, lat0)` to `(lon1, lat1)`.
fn road(id: i64, a: (f64, f64), b: (f64, f64), highway: &str) -> Way {
    let t = vec![("highway".to_string(), highway.to_string())];
    Way::from_lonlat_with(id, &[a, b], &t, &facett_map::MapPalette::cartographic(false))
        .expect("a 2-vertex line is a way")
}

/// **The T-junction fixture** — a residential running across the pane and a stub of
/// `class` coming down from the top and stopping ON its centreline, so the junction sits
/// at the pane centre. `stub` selects which of the three ways are in the scene, so the
/// same camera renders "the through road alone", "the stub alone" and "both".
///
/// `split` puts the two ways in **different tiles**, which is the arrangement the merge
/// actually has to survive: inside one tile the casing pass for the whole chunk precedes
/// the fill pass and the order alone would carry it, so a green there could be the chunk
/// order rather than the rank. Across tiles the stub's casing is submitted AFTER the
/// through road's fill, and only `LessEqual` against an equal `depth_of(rank)` can reject
/// it. Both arrangements are asserted.
fn t_junction(class: &str, which: &str, split: bool) -> MapFacet {
    let mut through = road(1, (18.040, 59.330), (18.080, 59.330), "residential");
    let mut stub = road(2, (18.060, 59.340), (18.060, 59.330), class);
    if split {
        // Tile 1 is submitted first whatever it holds, so the stub in tile 2 has its
        // casing drawn after the through road's fill is already in the depth buffer.
        through.tile_key = 1;
        stub.tile_key = 2;
    }
    let ways = match which {
        "through" => vec![through],
        "stub" => vec![stub],
        _ => vec![through, stub],
    };
    let view = MapView::new(LatLon::new(59.330, 18.060), 15.0, (CANVAS.x, CANVAS.y));
    MapFacet::new("gpu-t-junction", view, MapTheme::default())
        .with_layer(Box::new(WaysLayer::new("osm-ways", ways)))
}

/// sRGB luma of a read-back pixel.
fn luma8(p: [u8; 4]) -> f32 {
    0.2126 * f32::from(p[0]) + 0.7152 * f32::from(p[1]) + 0.0722 * f32::from(p[2])
}

/// One scan across the junction: the row through the through-road's centreline, for the
/// `±half` columns either side of the pane centre.
fn throat_row(img: &image::RgbaImage, half: i32) -> Vec<[u8; 4]> {
    let cy = img.height() / 2;
    let cx = i32::try_from(img.width()).expect("pane width fits an i32") / 2;
    (-half..=half)
        .map(|d| img.get_pixel(cx.saturating_add(d).clamp(0, i32::try_from(img.width() - 1).unwrap()) as u32, cy).0)
        .collect()
}

/// **A T-JUNCTION OF TWO ROADS OF ONE CLASS HAS NO CASING BAR ACROSS ITS THROAT** —
/// lane GG #26, the audit's red oracle, and the row where the audit turns out to be
/// **describing a defect the lane does not have**.
///
/// # What the audit says
///
/// > *casing = second draw of the same geometry per way (line.wgsl:126,170); casings of
/// > one class do not merge at intersections (Garmin does). Casing pass for the whole
/// > class before fills (stencil Replace) in `record_draws`.*
///
/// # What the lane actually does, and why the stencil pass is not needed
///
/// The casing is not a second draw **per way** — `LineRenderer::draw_indirect` binds the
/// casing uniform once and draws the whole CHUNK, then binds the fill uniform and draws
/// the same chunk again. Every casing in a chunk is already laid down before any fill in
/// it. Across chunks and tiles the merge is carried by Z instead: since MP-a a casing
/// rides at `depth_of(rank) + CASING_DEPTH_BIAS` and its fill at `depth_of(rank)`, and
/// the compare is `LessEqual` (GG #3). Two ways of ONE class have one rank, so a later
/// way's casing is **depth-rejected** exactly where any earlier way's fill of that class
/// already wrote — which is precisely "the casings of a class merge". A whole-class
/// stencil pass would buy the same picture at the cost of a second pass and of breaking
/// the casing ⊇ fill sample-mask nesting GG #2 rests on.
///
/// So this test asserts the property in the audit's own terms and finds it already held.
/// It is kept, and kept RED-armed, because the property is the one the owner looks at and
/// nothing else in the tree pinned it.
///
/// # The red arm is the CLASS axis, not a mutation
///
/// Replacing the stub with a MOTORWAY must put a casing bar across the residential — a
/// motorway outranks it, and a higher class cutting a lower one is the same Garmin rule
/// seen from the other side. If both arms read zero the metric is blind; if both read a
/// bar the merge is not happening. Only the pair means anything.
#[test]
fn casings_of_one_class_merge_at_a_t_junction_and_a_higher_class_still_cuts() {
    if !gated("casings_of_one_class_merge_at_a_t_junction_and_a_higher_class_still_cuts") {
        return;
    }
    const HALF: i32 = 26;
    // How much darker than the through road's own ink counts as casing. The casing is
    // near-black navy against a light road, so the gap is tens of luma steps.
    const BAR_DROP: f32 = 12.0;

    let shot = |class: &str, which: &str, split: bool| -> (u64, image::RgbaImage) {
        let (frames, img) = drive(t_junction(class, which, split), 3);
        let tiles = if split { "twotiles" } else { "onetile" };
        let path = format!("{}/scratch/cargo/gg26_t_{class}_{which}_{tiles}.png", std::env::var("HOME").unwrap_or_default());
        if image::DynamicImage::ImageRgba8(img.clone()).save(&path).is_ok() {
            eprintln!("LOOK: saved {path}");
        }
        (frames, img)
    };

    // The through road alone is the reference: what the junction row looks like with
    // nothing crossing it.
    let (f_ref, alone) = shot("residential", "through", false);
    let reference = throat_row(&alone, HALF);
    assert!(f_ref > 0, "the reference render must come off the GPU lane, not the CPU painter");
    let road_luma = luma8(reference[0]);
    // NON-VACUITY: the through road really is under the scan row, at its own colour and
    // not the paper.
    let paper = px_at(&alone, 0.95, 0.95);
    assert!(
        (road_luma - luma8(paper)).abs() > 8.0,
        "the scan row reads {road_luma:.1} luma and the paper reads {:.1} — the through \
         road is not on the row this test scans, so nothing below measures a junction",
        luma8(paper)
    );

    // How many columns of the throat go DARK when the stub is added, not counting the
    // columns the stub's own FILL legitimately owns (read off the stub rendered alone).
    let bar_for = |class: &str, split: bool| -> (usize, usize) {
        let (f_stub, stub_only) = shot(class, "stub", split);
        let (f_both, both) = shot(class, "both", split);
        assert!(f_stub > 0 && f_both > 0, "{class}: both renders must come off the GPU lane");
        let stub_row = throat_row(&stub_only, HALF);
        let both_row = throat_row(&both, HALF);
        let stub_paper = luma8(px_at(&stub_only, 0.95, 0.95));
        let mut bar = 0usize;
        let mut stub_ink = 0usize;
        let mut profile = String::new();
        for i in 0..reference.len() {
            // A column the stub's own ink does not reach at all is not part of the
            // junction and cannot be a bar.
            let stub_here = (luma8(stub_row[i]) - stub_paper).abs() > 4.0;
            if !stub_here {
                continue;
            }
            stub_ink += 1;
            // The stub's FILL owns this column — its own bright core, at or above the
            // through road's own luma. Nothing to claim there.
            if luma8(stub_row[i]) >= road_luma - BAR_DROP {
                continue;
            }
            // The stub is DARK here (its casing) and the through road was bright. If the
            // combined frame is dark too, the casing painted over the fill: a bar.
            if luma8(both_row[i]) < road_luma - BAR_DROP {
                bar += 1;
                profile.push('#');
            } else {
                profile.push('.');
            }
        }
        eprintln!(
            "[gg26] {class} stub, {} tile(s): {stub_ink} of {} throat columns carry stub ink, \
             {bar} of them go dark over the through road (# = bar, . = merged): {profile}",
            if split { 2 } else { 1 },
            reference.len()
        );
        (bar, stub_ink)
    };

    let (same_class_bar, same_ink) = bar_for("residential", false);
    let (split_bar, split_ink) = bar_for("residential", true);
    let (higher_class_bar, higher_ink) = bar_for("motorway", false);

    assert!(
        same_ink >= 6 && higher_ink >= 6 && split_ink >= 6,
        "the stub must actually cross the scan row in every arm ({same_ink} / {split_ink} / \
         {higher_ink} columns of ink) — a stub that misses the row proves nothing either way"
    );
    // THE CLAIM (audit #26): one class, no bar — in one tile, where the chunk's casing
    // pass precedes its fill pass, and across two, where only the equal rank can carry it.
    assert_eq!(
        same_class_bar, 0,
        "a T-junction of two residentials in ONE tile shows a {same_class_bar}-column casing \
         bar across its throat — the casings of one class are not merging"
    );
    assert_eq!(
        split_bar, 0,
        "a T-junction of two residentials in TWO TILES shows a {split_bar}-column casing bar \
         across its throat. Inside a tile the merge could be the chunk's casing-then-fill \
         order; across tiles the stub's casing is submitted after the through road's fill and \
         only `LessEqual` against an equal `depth_of(rank)` rejects it — so this is the arm \
         that says the merge is carried by the RANK"
    );
    // THE ANTI-BLIND-GREEN: the same metric on the same row DOES see a bar when the
    // classes differ, so a zero above is a merge and not a dead measurement.
    assert!(
        higher_class_bar >= 2,
        "MUTATION FAILED: a MOTORWAY stub laid only {higher_class_bar} dark columns across \
         the residential either. A higher class must cut a lower one, so a metric that \
         cannot see that cannot see a bar at all — and the zero above means nothing"
    );

    nornir_testmatrix::functional_status(
        "facett-map",
        "gpu_lane_casings_merge_per_class",
        same_class_bar == 0 && split_bar == 0 && higher_class_bar >= 2,
        &format!(
            "GG #26: residential stub lays {same_class_bar} casing columns across a residential \
             in one tile ({same_ink} throat columns) and {split_bar} across two ({split_ink}); \
             motorway stub lays {higher_class_bar} ({higher_ink})"
        ),
    );
}