facett-map/tests/tile_gpu_cache.rs
//! **The TILE lane's VRAM cache is REAL: a key uploaded once survives frames that
//! carry no payload, and the LRU evicts the tile nobody drew — never the ones being
//! drawn.**
//!
//! This is shit-list item 1 (retained per-tile GPU meshes): korp rebuilds its
//! `MapFacet` on every viewport apply, and the legacy whole-payload path re-baked +
//! re-uploaded EVERYTHING each time. The tile lane keys chunk sets by CONTENT in
//! `GpuMapRenderer::tile_gpu` — renderer state, which lives in egui's
//! `CallbackResources` and survives facet rebuilds — so a pan re-uploads only the
//! genuinely new ground and a zoom back onto cached ground uploads NOTHING.
//!
//! The observables are the renderer's own residency census (`tile_resident`,
//! `tile_cache_len`, `tile_cache_bytes`) and the known/wanted registry the facet's
//! hand-over decisions read (`tile_payload_needed`) — the same applied-output
//! posture as `upload_replaces_geometry.rs`.
//!
//! Run: `cargo test -p facett-map --features wgpu --test tile_gpu_cache`
#![cfg(feature = "wgpu")]
use facett_map::Way;
use facett_map::gpu::{GpuGeometry, GpuMapRenderer, build_gpu_geometry, ways_origin};
mod common;
fn roads(n: usize, first_id: i64, lon0: f64) -> Vec<Way> {
let c = egui::Color32::from_rgb(0xE8, 0xB0, 0x30);
(0..n)
.map(|i| {
let lon = lon0 + i as f64 * 0.1;
Way::line(first_id + i as i64, "road", vec![(lon, 47.0), (lon + 0.05, 47.02), (lon + 0.09, 47.0)], c, 5.0)
})
.collect()
}
fn geometry(ways: &[Way]) -> GpuGeometry {
let origin = ways_origin(ways.iter()).expect("non-empty fixture has an origin");
build_gpu_geometry(ways.iter().map(|w| (w, w.style.lod as u32)), origin)
}
/// **RED-WHEN-BROKEN.** Upload two keyed tiles, then ask the residency + registry
/// questions a later frame's hand-over decision turns on. Deleting the cache (or the
/// `tile_registry_mark_known` on upload) flips `tile_payload_needed` back to `true`
/// for a resident key — which is the legacy re-upload-every-apply behaviour this
/// lane deletes.
#[test]
fn an_uploaded_tile_stays_resident_and_stops_asking_for_its_payload() {
let Some((device, queue)) = common::device_or_skip(
"an_uploaded_tile_stays_resident_and_stops_asking_for_its_payload",
"a tile uploaded once is served from VRAM on every later frame",
) else {
return;
};
let fmt = wgpu::TextureFormat::Rgba8UnormSrgb;
let mut r = GpuMapRenderer::new(&device, fmt);
let (key_a, key_b) = (0xA11CE_u128, 0xB0B_u128);
assert!(
facett_map::gpu::tile_payload_needed(key_a),
"an unknown key MUST ask for its payload — this is the first-sight upload",
);
r.upload_tile_for_test(&device, &queue, key_a, &geometry(&roads(40, 1, 9.0)));
r.upload_tile_for_test(&device, &queue, key_b, &geometry(&roads(40, 100, 19.0)));
assert!(r.tile_resident(key_a) && r.tile_resident(key_b), "both tiles are resident");
assert!(r.tile_cache_bytes() > 0, "the census sees real VRAM bytes");
assert!(
!facett_map::gpu::tile_payload_needed(key_a),
"a RESIDENT key must NOT ask for its payload — asking is the re-upload-per-apply \
regression this lane exists to delete",
);
assert!(!facett_map::gpu::tile_payload_needed(key_b), "same for the second tile");
}
/// **The LRU evicts the tile nobody drew — and eviction re-opens the payload ask.**
/// Three tiles under a budget that fits two: the two in the current draw set
/// survive, the third falls, and `tile_payload_needed` for the victim flips back to
/// `true` (the self-healing re-send), while the survivors still answer `false`.
#[test]
fn eviction_takes_the_undrawn_tile_and_reopens_its_payload_ask() {
let Some((device, queue)) = common::device_or_skip(
"eviction_takes_the_undrawn_tile_and_reopens_its_payload_ask",
"the tile LRU evicts outside the current draw set and re-asks",
) else {
return;
};
let fmt = wgpu::TextureFormat::Rgba8UnormSrgb;
let mut r = GpuMapRenderer::new(&device, fmt);
let (ka, kb, kc) = (0x1_u128, 0x2_u128, 0x3_u128);
r.upload_tile_for_test(&device, &queue, ka, &geometry(&roads(40, 1, 9.0)));
r.upload_tile_for_test(&device, &queue, kb, &geometry(&roads(40, 100, 19.0)));
r.upload_tile_for_test(&device, &queue, kc, &geometry(&roads(40, 200, 29.0)));
assert_eq!(r.tile_cache_len(), 3, "precondition: three resident tiles");
let per_tile = r.tile_cache_bytes() / 3;
// Budget for two: drawing {a, c} must evict b — the one NOT in the draw set —
// regardless of upload order.
r.set_tile_vram_budget_for_test(per_tile * 2 + per_tile / 2);
r.evict_tiles_for_test(&[ka, kc]);
assert!(r.tile_resident(ka), "a tile IN the draw set survives");
assert!(r.tile_resident(kc), "the other drawn tile survives too");
assert!(!r.tile_resident(kb), "the tile NOBODY drew is the victim");
assert!(
facett_map::gpu::tile_payload_needed(kb),
"the evicted key must ASK again — an eviction that keeps claiming residency \
is a permanently blank tile",
);
assert!(
!facett_map::gpu::tile_payload_needed(ka),
"and the survivor still does not ask",
);
}
/// **THE TILE LANE ACTUALLY CULLS — and the count it hands `draw_indirect` is right.**
///
/// The gap this closes is a hole in this very file. Everything above asserts residency,
/// bytes and the wanted-registry, all of which are true of a lane that draws NOTHING —
/// and on 2026-08-22 that is exactly what the lane briefly was: the ordered-compaction
/// change (`cull.wgsl`'s three passes) landed on `run_cull` and left
/// `run_cull_tiles_clocked` on the bare copy pass, so `main` read offsets `count_pass`
/// had never written and `draw_count` stayed at the CPU's reset of zero. The whole
/// suite was green. Nothing was reading a tile's `vertex_count`.
///
/// The applied output of a cull is that number, so this reads it, out of the very
/// buffer `draw_indirect` reads, after the real dispatch.
///
/// RED: point `run_cull_tiles_clocked` at `cull_pipeline` alone (the pre-2026-08-22
/// single dispatch) — every count comes back 0.
#[test]
fn the_tile_lane_culls_and_hands_the_draw_a_real_count() {
let Some((device, queue)) = common::device_or_skip(
"the_tile_lane_culls_and_hands_the_draw_a_real_count",
"the tile lane's cull produces a real, viewport-tracking vertex count",
) else {
return;
};
let ways = roads(40, 1, 9.0);
let origin = ways_origin(ways.iter()).expect("non-empty fixture has an origin");
let geo = geometry(&ways);
let total: u32 = geo.line_meta.iter().map(|m| m.vert_count).sum();
assert!(total > 0, "the fixture must bake quads at all");
let mut r = GpuMapRenderer::new(&device, wgpu::TextureFormat::Rgba8Unorm);
// A key OF ITS OWN. The known/wanted registry `tile_payload_needed` reads is
// PROCESS-global, so two `#[test]`s sharing a key share state across threads:
// reusing `0xA11CE` here made `an_uploaded_tile_stays_resident_..` fail about one
// run in three, on its "an unknown key MUST ask for its payload" line.
let key = 0xC0FFEE_u128;
r.upload_tile_for_test(&device, &queue, key, &geo);
assert!(r.tile_resident(key), "the tile is in VRAM");
// The tile's own rebased cull space — the same rebasing `MapTransform::viewport_bbox`
// hands the lane per tile.
let bbox = |lon0: f64, lon1: f64| {
let (x0, y0) = facett_map::mercator(lon0, 46.0);
let (x1, y1) = facett_map::mercator(lon1, 48.0);
[
(x0 - origin[0]) as f32,
(y0.min(y1) - origin[1]) as f32,
(x1 - origin[0]) as f32,
(y0.max(y1) - origin[1]) as f32,
]
};
// ── Everything visible ────────────────────────────────────────────────────
r.cull_tiles_for_test(&device, &queue, &[(key, bbox(0.0, 30.0))], 3, 0.0);
let (_, quads_all) = r.tile_draw_counts(&device, &queue, key);
let all: u32 = quads_all.iter().sum();
assert_eq!(
all, total,
"a viewport containing the whole tile must draw every one of its {total} quad vertices, \
got {all}. Zero here means the tile lane's cull never ran its count/scan passes."
);
// ── Nothing visible: the same lane must go to zero ────────────────────────
r.cull_tiles_for_test(&device, &queue, &[(key, bbox(60.0, 61.0))], 3, 0.0);
let (_, quads_none) = r.tile_draw_counts(&device, &queue, key);
let none: u32 = quads_none.iter().sum();
assert_eq!(none, 0, "an empty viewport must draw NOTHING — got {none} of {total}");
// ── A slice: strictly between, so neither arm above can be passing vacuously
r.cull_tiles_for_test(&device, &queue, &[(key, bbox(9.0, 11.0))], 3, 0.0);
let (_, quads_part) = r.tile_draw_counts(&device, &queue, key);
let part: u32 = quads_part.iter().sum();
assert!(
part > 0 && part < all,
"a partial viewport must draw SOME of the tile: got {part}, all is {all}"
);
eprintln!("[tile-cull] all={all}/{total} none={none} partial={part}");
nornir_testmatrix::functional_status(
"facett-map",
"tile_lane_culls_for_real",
all == total && none == 0 && part > 0 && part < all,
&format!("tile quad vertex_count: all={all}/{total}, empty={none}, partial={part}"),
);
}
/// **RED-WHEN-BROKEN — lane GG #13, "cull dispatch at rest".** The tile lane used to
/// dispatch its three cull passes over every stream of every current tile on EVERY
/// frame — camera moving or not — after resetting every chunk's indirect counter and
/// rewriting every tile's cull viewport and five draw uniforms. On a still frame every
/// one of those writes is byte-identical to the previous frame's, and the cull it
/// feeds lands on the same `vertex_count`s. So: with the draw set, the cull viewport
/// and the upload generation all unchanged, the renderer must record **zero** cull
/// passes, and the previous frame's indirect buffers must still hold the draw.
///
/// The three arms after the still run are what keep the skip from being a dead cull:
/// a viewport change dispatches (and the count moves), moving back dispatches again
/// (the result is recomputed, not stale), and a fresh tile entering the draw set
/// dispatches (its indirect counter was never written — a skip there would draw
/// nothing from it).
#[test]
fn a_still_frame_dispatches_no_cull_and_a_moved_or_uploaded_one_does() {
let Some((device, queue)) = common::device_or_skip(
"a_still_frame_dispatches_no_cull_and_a_moved_or_uploaded_one_does",
"100 frames at rest record 0 cull dispatches; a move or an upload records one",
) else {
return;
};
const STILL_FRAMES: u64 = 100;
let ways = roads(40, 2001, 9.0);
let origin = ways_origin(ways.iter()).expect("non-empty fixture has an origin");
let geo = geometry(&ways);
let total: u32 = geo.line_meta.iter().map(|m| m.vert_count).sum();
assert!(total > 0, "the fixture must bake quads at all");
let mut r = GpuMapRenderer::new(&device, wgpu::TextureFormat::Rgba8Unorm);
// Keys of this test's OWN — the known/wanted registry is process-global (see the
// sibling test's note).
let key = 0x5711_u128;
let key2 = 0x5712_u128;
r.upload_tile_for_test(&device, &queue, key, &geo);
let bbox = |lon0: f64, lon1: f64| {
let (x0, y0) = facett_map::mercator(lon0, 46.0);
let (x1, y1) = facett_map::mercator(lon1, 48.0);
[
(x0 - origin[0]) as f32,
(y0.min(y1) - origin[1]) as f32,
(x1 - origin[0]) as f32,
(y0.max(y1) - origin[1]) as f32,
]
};
let vp_all = bbox(0.0, 30.0);
let vp_none = bbox(60.0, 61.0);
let quads = |r: &GpuMapRenderer, k: u128| -> u32 { r.tile_draw_counts(&device, &queue, k).1.iter().sum() };
// ── Frame 1: the camera arrives. This one MUST cull. ─────────────────────
r.cull_tiles_for_test(&device, &queue, &[(key, vp_all)], 3, 0.0);
assert_eq!(r.cull_dispatch_count(), 1, "the first frame over a fresh tile culls exactly once");
let all = quads(&r, key);
assert_eq!(all, total, "the first frame draws the whole tile");
// ── 100 still frames: same draw set, same viewport, nothing uploaded. ─────
let t0 = std::time::Instant::now();
for _ in 0..STILL_FRAMES {
r.cull_tiles_for_test(&device, &queue, &[(key, vp_all)], 3, 0.0);
}
let still_us = t0.elapsed().as_micros();
let redone = r.cull_dispatch_count() - 1;
eprintln!(
"[cull-at-rest t14s/890M] {STILL_FRAMES} still frames: {redone} cull dispatches, {still_us} us total \
({:.1} us/frame incl. submit+poll)",
still_us as f64 / STILL_FRAMES as f64
);
assert_eq!(
redone, 0,
"{STILL_FRAMES} frames with the camera at rest re-dispatched the cull {redone} times; \
with cull_vp unchanged and nothing uploaded the count must not move"
);
assert_eq!(
quads(&r, key),
all,
"a skipped cull must leave the previous frame's indirect count intact — the draw reads it"
);
// ── The camera moves off the tile: the cull runs, and the count follows. ─
r.cull_tiles_for_test(&device, &queue, &[(key, vp_none)], 3, 0.0);
assert_eq!(r.cull_dispatch_count(), 2, "a viewport change must dispatch the cull again");
assert_eq!(quads(&r, key), 0, "off-tile viewport draws nothing");
for _ in 0..STILL_FRAMES {
r.cull_tiles_for_test(&device, &queue, &[(key, vp_none)], 3, 0.0);
}
assert_eq!(r.cull_dispatch_count(), 2, "still frames after the move stay skipped");
// ── Back onto the tile: recomputed, not served stale. ─────────────────────
r.cull_tiles_for_test(&device, &queue, &[(key, vp_all)], 3, 0.0);
assert_eq!(r.cull_dispatch_count(), 3, "moving back must dispatch — the skip is not a memo of old results");
assert_eq!(quads(&r, key), all, "moving back redraws the whole tile");
// ── The LOD changes under the same bbox: it is part of cull_vp, so it culls. ─
r.cull_tiles_for_test(&device, &queue, &[(key, vp_all)], 4, 0.0);
assert_eq!(r.cull_dispatch_count(), 4, "an lod change is a cull_vp change and must dispatch");
// ── A fresh tile enters the draw set: its counter was never written. ──────
r.upload_tile_for_test(&device, &queue, key2, &geo);
r.cull_tiles_for_test(&device, &queue, &[(key, vp_all), (key2, vp_all)], 4, 0.0);
assert_eq!(r.cull_dispatch_count(), 5, "a fresh upload in the draw set must dispatch");
assert_eq!(quads(&r, key2), total, "the new tile is culled for real, not left at zero");
for _ in 0..STILL_FRAMES {
r.cull_tiles_for_test(&device, &queue, &[(key, vp_all), (key2, vp_all)], 4, 0.0);
}
assert_eq!(r.cull_dispatch_count(), 5, "two tiles at rest are as still as one");
nornir_testmatrix::functional_status(
"facett-map",
"cull_dispatch_at_rest",
redone == 0 && r.cull_dispatch_count() == 5,
&format!("{STILL_FRAMES} still frames: {redone} cull dispatches ({still_us} us); moves/uploads: 5 total"),
);
}