Module cuda_backend 

2026-09-25: The CUDA GpuBackend: MetraleCudaBackend, the driver API declarations its submodules call, and its device-allocation ledger.

Owner: gpu-runtime (CUDA backend). Invariants:

• A pointer that alloc or alloc_managed returns is on live_allocs until free or sweep_unreleased takes it off; free takes it off before calling cuMemFree_v2, whatever that call returns.
• alloc_arena allocations are never on live_allocs.

Modules

arch_preflight

2026-09-25: Refuse a GPU whose compute capability cannot run the kernels’ architecture, before MetraleCudaBackend::new loads any module, with an error that names both ([metrale_core::arch::ArchMismatch]); also warn when the device’s SM count differs from the target’s sm_count.

tensormap

2026-09-25: TMA descriptors (CUtensorMap) for cp.async.bulk.tensor loads, encoded with the driver’s cuTensorMapEncodeTiled.

Structs

LedgerProbe

2026-10-01: A read handle on one backend’s ledger, for a reader that cannot hold the backend, which moves into the model it builds. It reads the map MetraleCudaBackend::live_bytes reads, so the two always agree.

MetraleCudaBackend

2026-09-25: The CUDA GpuBackend, holding one model’s kernel modules.

Functions

cuda_free_memory_bytes

2026-09-25: Free device memory without a GpuBackend, for the OOM watchdog and the TUI. None when cuMemGetInfo_v2 fails.

spawn_oom_watchdog

2026-09-25: Spawn the OOM watchdog on the tokio runtime, or return None if one was already spawned.