gitadmin/foxhunt
1
0
Fork 0
Code Issues Pull Requests Actions Packages Projects Releases Wiki Activity

diag(ml-backtesting): finer NaN instrumentation — per-vwap-write-site + last-bad-vwap capture

Browse Source 
v1 instrumentation (S1.20) eliminated 3 of 6 hypotheses: apply_fill_to_pos
arithmetic is fully clean (avg_px=0 realised=0 realized_pnl=0). But
pnl_track open branch still saw vwap_entry=0 (194 times) or > 21M (216
times) at the open transition.

v2 adds per-vwap-write-site counters so we can pinpoint which apply_fill
branch produced the bad vwap, plus captures the actual last-bad-vwap
value and the path id (1..6 = open-flat, scale-in zero/huge, flip-beyond
zero/huge, session-gap-saw-stale).

The 7th counter (vwap_session_gap_was_bad) fires when the session-gap
force-close path sees pos.vwap_entry already in a corrupt state pre-gap.

Logged at end of smoke as `nan_counters_v2: zero_flat=N huge_flat=N
zero_scale=N huge_scale=N zero_flip=N huge_flip=N session_gap_was_bad=N
| b0_last_bad_vwap=X b0_last_bad_path=N`.

19/19 stop_controller CUDA tests pass; 5/5 decision_floor_coldstart pass.

Co-Authored-By: Claude Opus 4.7 (1M context) <noreply@anthropic.com>
This commit is contained in:
jgrusewski
2026-05-20 13:27:03 +02:00
parent 025554afcd
commit fc41440dc9
4 changed files with 246 additions and 5 deletions
Download Patch File Download Diff File Expand all files Collapse all files

86
crates/ml-backtesting/cuda/resting_orders.cu
View File

@@ -105,6 +105,12 @@ __device__ static void cancel_oco_pair(Orders& orders, unsigned char oco_pair) {
// nan_realised — incremented when realised P&L in counter-direction branch is NaN/Inf.
// nan_realized_pnl — incremented when pos.realized_pnl becomes non-finite after += or -=.
// On NaN-producing events the function returns early WITHOUT poisoning pos state.
//
// S1.21: per-vwap-write-site counters.
// vwap_zero_open_flat / vwap_huge_open_flat — open-from-flat branch (line 131).
// vwap_zero_scale_in / vwap_huge_scale_in — scale-in branch (line 136).
// vwap_zero_flip / vwap_huge_flip — flip-beyond branch (line 150).
// last_bad_vwap / last_bad_path — capture site + value.
__device__ static void apply_fill_to_pos(
Pos& pos,
float filled_lots,
@@ -115,7 +121,15 @@ __device__ static void apply_fill_to_pos(
float* __restrict__ total_fees_per_b,
unsigned int* __restrict__ nan_avg_px,
unsigned int* __restrict__ nan_realised,
unsigned int* __restrict__ nan_realized_pnl
unsigned int* __restrict__ nan_realized_pnl,
unsigned int* __restrict__ vwap_zero_open_flat,
unsigned int* __restrict__ vwap_huge_open_flat,
unsigned int* __restrict__ vwap_zero_scale_in,
unsigned int* __restrict__ vwap_huge_scale_in,
unsigned int* __restrict__ vwap_zero_flip,
unsigned int* __restrict__ vwap_huge_flip,
float* __restrict__ last_bad_vwap,
unsigned int* __restrict__ last_bad_path
) {
if (filled_lots <= 0.0f) return;
const float avg_px = total_cost / filled_lots;
@@ -129,11 +143,31 @@ __device__ static void apply_fill_to_pos(
if (pos.position_lots == 0) {
pos.vwap_entry = avg_px;
// S1.21: open-from-flat write site.
if (pos.vwap_entry == 0.0f) {
vwap_zero_open_flat[b] += 1u;
last_bad_vwap[b] = 0.0f;
last_bad_path[b] = 1u;
} else if (!isfinite(pos.vwap_entry) || fabsf(pos.vwap_entry) > 21000000.0f) {
vwap_huge_open_flat[b] += 1u;
last_bad_vwap[b] = pos.vwap_entry;
last_bad_path[b] = 1u;
}
} else if ((prev_pos_f > 0.0f) == (sgn > 0.0f)) {
const float prev_notional = prev_pos_f * pos.vwap_entry;
const float new_notional = prev_notional + sgn * total_cost;
const float new_abs = (new_pos_f > 0.0f) ? new_pos_f : -new_pos_f;
pos.vwap_entry = (new_abs > 0.0f) ? (new_notional / (sgn * new_abs)) : 0.0f;
// S1.21: scale-in write site.
if (pos.vwap_entry == 0.0f) {
vwap_zero_scale_in[b] += 1u;
last_bad_vwap[b] = 0.0f;
last_bad_path[b] = 2u;
} else if (!isfinite(pos.vwap_entry) || fabsf(pos.vwap_entry) > 21000000.0f) {
vwap_huge_scale_in[b] += 1u;
last_bad_vwap[b] = pos.vwap_entry;
last_bad_path[b] = 3u;
}
} else {
const float prev_abs = (prev_pos_f > 0.0f) ? prev_pos_f : -prev_pos_f;
const float unwind = (filled_lots < prev_abs) ? filled_lots : prev_abs;
@@ -147,7 +181,19 @@ __device__ static void apply_fill_to_pos(
if (!isfinite(pos.realized_pnl)) {
nan_realized_pnl[b] += 1u;
}
if (filled_lots > prev_abs) { pos.vwap_entry = avg_px; }
if (filled_lots > prev_abs) {
pos.vwap_entry = avg_px;
// S1.21: flip-beyond write site.
if (pos.vwap_entry == 0.0f) {
vwap_zero_flip[b] += 1u;
last_bad_vwap[b] = 0.0f;
last_bad_path[b] = 4u;
} else if (!isfinite(pos.vwap_entry) || fabsf(pos.vwap_entry) > 21000000.0f) {
vwap_huge_flip[b] += 1u;
last_bad_vwap[b] = pos.vwap_entry;
last_bad_path[b] = 5u;
}
}
}
pos.position_lots = (int)new_pos_f;
@@ -193,6 +239,16 @@ extern "C" __global__ void resting_orders_step(
unsigned int* __restrict__ nan_avg_px,
unsigned int* __restrict__ nan_realised,
unsigned int* __restrict__ nan_realized_pnl,
// S1.21: per-vwap-write-site counters + last-bad-vwap capture.
unsigned int* __restrict__ vwap_zero_open_flat,
unsigned int* __restrict__ vwap_huge_open_flat,
unsigned int* __restrict__ vwap_zero_scale_in,
unsigned int* __restrict__ vwap_huge_scale_in,
unsigned int* __restrict__ vwap_zero_flip,
unsigned int* __restrict__ vwap_huge_flip,
unsigned int* __restrict__ vwap_session_gap_was_bad,
float* __restrict__ last_bad_vwap,
unsigned int* __restrict__ last_bad_path,
int n_backtests
) {
int b = blockIdx.x;
@@ -217,6 +273,14 @@ extern "C" __global__ void resting_orders_step(
// transition via open_trade_state scratch and emit a close TradeRecord.
// No synthetic P&L is added — records show realised_pnl from whatever
// was accumulated before the gap (honest: cannot fill across a halt).
// S1.21: capture pre-close vwap_entry corruption if already bad.
if (pos.vwap_entry == 0.0f || !isfinite(pos.vwap_entry)
|| fabsf(pos.vwap_entry) > 21000000.0f)
{
vwap_session_gap_was_bad[b] += 1u;
last_bad_vwap[b] = pos.vwap_entry;
last_bad_path[b] = 6u;
}
pos.position_lots = 0;
}
last_event_ts[b] = current_ts_ns;
@@ -260,7 +324,11 @@ extern "C" __global__ void resting_orders_step(
const float cost = take * s.price;
apply_fill_to_pos(pos, take, cost, s.side,
b, cost_per_lot_per_side_per_b, total_fees_per_b,
nan_avg_px, nan_realised, nan_realized_pnl);
nan_avg_px, nan_realised, nan_realized_pnl,
vwap_zero_open_flat, vwap_huge_open_flat,
vwap_zero_scale_in, vwap_huge_scale_in,
vwap_zero_flip, vwap_huge_flip,
last_bad_vwap, last_bad_path);
s.size_remaining -= take;
emit_audit(audit_base, audit_head, b, (unsigned int)ring_cap_events,
current_ts_ns,
@@ -291,7 +359,11 @@ extern "C" __global__ void resting_orders_step(
if (filled > 0.0f) {
apply_fill_to_pos(pos, filled, cost, s.side,
b, cost_per_lot_per_side_per_b, total_fees_per_b,
nan_avg_px, nan_realised, nan_realized_pnl);
nan_avg_px, nan_realised, nan_realized_pnl,
vwap_zero_open_flat, vwap_huge_open_flat,
vwap_zero_scale_in, vwap_huge_scale_in,
vwap_zero_flip, vwap_huge_flip,
last_bad_vwap, last_bad_path);
s.size_remaining -= filled;
emit_audit(audit_base, audit_head, b, (unsigned int)ring_cap_events,
current_ts_ns,
@@ -338,7 +410,11 @@ extern "C" __global__ void resting_orders_step(
if (filled > 0.0f) {
apply_fill_to_pos(pos, filled, cost, st.side,
b, cost_per_lot_per_side_per_b, total_fees_per_b,
nan_avg_px, nan_realised, nan_realized_pnl);
nan_avg_px, nan_realised, nan_realized_pnl,
vwap_zero_open_flat, vwap_huge_open_flat,
vwap_zero_scale_in, vwap_huge_scale_in,
vwap_zero_flip, vwap_huge_flip,
last_bad_vwap, last_bad_path);
emit_audit(audit_base, audit_head, b, (unsigned int)ring_cap_events,
current_ts_ns,
pack_slot_tag(1 /*SlotKind::Stop*/, (unsigned char)j, st.gen_counter),

23
crates/ml-backtesting/src/harness.rs
View File

@@ -305,6 +305,29 @@ impl BacktestHarness {
Err(e) => eprintln!("nan_counters read failed: {e}"),
}
// S1.21: per-vwap-write-site counters. Printed after v1 line so both
// appear together in the smoke log.
match self.sim.read_nan_counters_v2() {
Ok(c) => {
let sum_u32 = |v: &[u32]| -> u64 { v.iter().map(|&x| x as u64).sum() };
let last_bad_vwap_b0 = c.last_bad_vwap.first().copied().unwrap_or(0.0);
let last_bad_path_b0 = c.last_bad_path.first().copied().unwrap_or(0);
eprintln!(
"nan_counters_v2: zero_flat={} huge_flat={} zero_scale={} huge_scale={} zero_flip={} huge_flip={} session_gap_was_bad={} | b0_last_bad_vwap={} b0_last_bad_path={}",
sum_u32(&c.vwap_zero_open_flat),
sum_u32(&c.vwap_huge_open_flat),
sum_u32(&c.vwap_zero_scale_in),
sum_u32(&c.vwap_huge_scale_in),
sum_u32(&c.vwap_zero_flip),
sum_u32(&c.vwap_huge_flip),
sum_u32(&c.vwap_session_gap_was_bad),
last_bad_vwap_b0,
last_bad_path_b0,
);
}
Err(e) => eprintln!("nan_counters_v2 read failed: {e}"),
}
Ok(RunStats {
events_processed: self.event_count,
decisions_taken: total_decisions,

119
crates/ml-backtesting/src/sim/mod.rs
View File

@@ -60,6 +60,40 @@ pub struct NanCounters {
pub defensive_exit_clamp: Vec<u32>,
}
/// Per-backtest per-vwap-write-site NaN instrumentation counters (S1.21).
///
/// Finer-grained than NanCounters: each counter corresponds to exactly one
/// vwap_entry write site inside apply_fill_to_pos (or the session-gap path).
/// Read via `LobSimCuda::read_nan_counters_v2`.
///
/// Path ids stored in last_bad_path:
/// 1 = open-from-flat (line 131)
/// 2 = scale-in zero (line 136)
/// 3 = scale-in huge (line 136)
/// 4 = flip-beyond zero (line 150)
/// 5 = flip-beyond huge (line 150)
/// 6 = session-gap saw already-corrupt vwap_entry
pub struct NanCountersV2 {
/// apply_fill_to_pos open-from-flat branch wrote vwap_entry == 0.
pub vwap_zero_open_flat: Vec<u32>,
/// apply_fill_to_pos open-from-flat branch wrote |vwap_entry| > 21M or non-finite.
pub vwap_huge_open_flat: Vec<u32>,
/// apply_fill_to_pos scale-in branch wrote vwap_entry == 0.
pub vwap_zero_scale_in: Vec<u32>,
/// apply_fill_to_pos scale-in branch wrote |vwap_entry| > 21M or non-finite.
pub vwap_huge_scale_in: Vec<u32>,
/// apply_fill_to_pos flip-beyond branch wrote vwap_entry == 0.
pub vwap_zero_flip: Vec<u32>,
/// apply_fill_to_pos flip-beyond branch wrote |vwap_entry| > 21M or non-finite.
pub vwap_huge_flip: Vec<u32>,
/// session-gap force-close saw pos.vwap_entry already corrupt (zero / non-finite / > 21M).
pub vwap_session_gap_was_bad: Vec<u32>,
/// Last bad vwap_entry value observed (per backtest; 0.0 if none).
pub last_bad_vwap: Vec<f32>,
/// Path id of the last bad vwap_entry write (per backtest; 0 if none).
pub last_bad_path: Vec<u32>,
}
pub struct LobSimCuda {
n_backtests: usize,
stream: Arc<CudaStream>,
@@ -162,6 +196,16 @@ pub struct LobSimCuda {
zero_vwap_at_open_d: CudaSlice<u32>, // [n_backtests]
saturated_vwap_at_open_d: CudaSlice<u32>, // [n_backtests]
defensive_exit_clamp_d: CudaSlice<u32>, // [n_backtests]
// S1.21: per-vwap-write-site counters + last-bad-vwap capture.
vwap_zero_open_flat_d: CudaSlice<u32>, // [n_backtests]
vwap_huge_open_flat_d: CudaSlice<u32>, // [n_backtests]
vwap_zero_scale_in_d: CudaSlice<u32>, // [n_backtests]
vwap_huge_scale_in_d: CudaSlice<u32>, // [n_backtests]
vwap_zero_flip_d: CudaSlice<u32>, // [n_backtests]
vwap_huge_flip_d: CudaSlice<u32>, // [n_backtests]
vwap_session_gap_was_bad_d: CudaSlice<u32>, // [n_backtests]
last_bad_vwap_d: CudaSlice<f32>, // [n_backtests]
last_bad_path_d: CudaSlice<u32>, // [n_backtests]
}
impl LobSimCuda {
@@ -366,6 +410,34 @@ impl LobSimCuda {
let defensive_exit_clamp_d = stream
.alloc_zeros::<u32>(n_backtests)
.context("alloc defensive_exit_clamp_d")?;
// S1.21: per-vwap-write-site counters + last-bad-vwap capture.
let vwap_zero_open_flat_d = stream
.alloc_zeros::<u32>(n_backtests)
.context("alloc vwap_zero_open_flat_d")?;
let vwap_huge_open_flat_d = stream
.alloc_zeros::<u32>(n_backtests)
.context("alloc vwap_huge_open_flat_d")?;
let vwap_zero_scale_in_d = stream
.alloc_zeros::<u32>(n_backtests)
.context("alloc vwap_zero_scale_in_d")?;
let vwap_huge_scale_in_d = stream
.alloc_zeros::<u32>(n_backtests)
.context("alloc vwap_huge_scale_in_d")?;
let vwap_zero_flip_d = stream
.alloc_zeros::<u32>(n_backtests)
.context("alloc vwap_zero_flip_d")?;
let vwap_huge_flip_d = stream
.alloc_zeros::<u32>(n_backtests)
.context("alloc vwap_huge_flip_d")?;
let vwap_session_gap_was_bad_d = stream
.alloc_zeros::<u32>(n_backtests)
.context("alloc vwap_session_gap_was_bad_d")?;
let last_bad_vwap_d = stream
.alloc_zeros::<f32>(n_backtests)
.context("alloc last_bad_vwap_d")?;
let last_bad_path_d = stream
.alloc_zeros::<u32>(n_backtests)
.context("alloc last_bad_path_d")?;
Ok(Self {
n_backtests,
@@ -430,6 +502,15 @@ impl LobSimCuda {
zero_vwap_at_open_d,
saturated_vwap_at_open_d,
defensive_exit_clamp_d,
vwap_zero_open_flat_d,
vwap_huge_open_flat_d,
vwap_zero_scale_in_d,
vwap_huge_scale_in_d,
vwap_zero_flip_d,
vwap_huge_flip_d,
vwap_session_gap_was_bad_d,
last_bad_vwap_d,
last_bad_path_d,
})
}
@@ -595,6 +676,35 @@ impl LobSimCuda {
})
}
/// S1.21: per-vwap-write-site counters.
///
/// All counters are zero at construction. Incremented in
/// `apply_fill_to_pos` at each vwap_entry write site and in the
/// session-gap force-close path in `resting_orders_step`.
pub fn read_nan_counters_v2(&self) -> Result<NanCountersV2> {
let read_u32 = |slot: &CudaSlice<u32>| -> Result<Vec<u32>> {
let mut buf = vec![0u32; self.n_backtests];
self.stream.memcpy_dtoh(slot, buf.as_mut_slice())?;
Ok(buf)
};
let read_f32 = |slot: &CudaSlice<f32>| -> Result<Vec<f32>> {
let mut buf = vec![0f32; self.n_backtests];
self.stream.memcpy_dtoh(slot, buf.as_mut_slice())?;
Ok(buf)
};
Ok(NanCountersV2 {
vwap_zero_open_flat: read_u32(&self.vwap_zero_open_flat_d)?,
vwap_huge_open_flat: read_u32(&self.vwap_huge_open_flat_d)?,
vwap_zero_scale_in: read_u32(&self.vwap_zero_scale_in_d)?,
vwap_huge_scale_in: read_u32(&self.vwap_huge_scale_in_d)?,
vwap_zero_flip: read_u32(&self.vwap_zero_flip_d)?,
vwap_huge_flip: read_u32(&self.vwap_huge_flip_d)?,
vwap_session_gap_was_bad: read_u32(&self.vwap_session_gap_was_bad_d)?,
last_bad_vwap: read_f32(&self.last_bad_vwap_d)?,
last_bad_path: read_u32(&self.last_bad_path_d)?,
})
}
/// Upload per-backtest price-range bounds. Call ONCE before the first
/// apply_snapshot. min=0.0 / max=f32::INFINITY disables the check for
/// that backtest (default after LobSimCuda::new).
@@ -1271,6 +1381,15 @@ impl LobSimCuda {
.arg(&mut self.nan_avg_px_d)
.arg(&mut self.nan_realised_d)
.arg(&mut self.nan_realized_pnl_d)
.arg(&mut self.vwap_zero_open_flat_d)
.arg(&mut self.vwap_huge_open_flat_d)
.arg(&mut self.vwap_zero_scale_in_d)
.arg(&mut self.vwap_huge_scale_in_d)
.arg(&mut self.vwap_zero_flip_d)
.arg(&mut self.vwap_huge_flip_d)
.arg(&mut self.vwap_session_gap_was_bad_d)
.arg(&mut self.last_bad_vwap_d)
.arg(&mut self.last_bad_path_d)
.arg(&n)
.launch(cfg)?;
}

23
crates/ml-backtesting/tests/stop_controller.rs
View File

@@ -1191,3 +1191,26 @@ fn price_range_rejection_skips_snapshot() -> Result<()> {
Ok(())
}
/// S1.21: All per-vwap-write-site v2 counters initialize to zero at
/// LobSimCuda construction; last_bad_path initializes to zero (no bad
/// write observed yet).
#[test]
#[ignore = "requires CUDA"]
fn nan_counters_v2_initialize_to_zero() -> Result<()> {
let dev = match MlDevice::cuda(0) {
Ok(d) => d,
Err(e) => { eprintln!("skipping: cuda unavailable ({e})"); return Ok(()); }
};
let sim = LobSimCuda::new(1, &dev)?;
let c = sim.read_nan_counters_v2()?;
assert_eq!(c.vwap_zero_open_flat[0], 0, "vwap_zero_open_flat must initialize to 0");
assert_eq!(c.vwap_huge_open_flat[0], 0, "vwap_huge_open_flat must initialize to 0");
assert_eq!(c.vwap_zero_scale_in[0], 0, "vwap_zero_scale_in must initialize to 0");
assert_eq!(c.vwap_huge_scale_in[0], 0, "vwap_huge_scale_in must initialize to 0");
assert_eq!(c.vwap_zero_flip[0], 0, "vwap_zero_flip must initialize to 0");
assert_eq!(c.vwap_huge_flip[0], 0, "vwap_huge_flip must initialize to 0");
assert_eq!(c.vwap_session_gap_was_bad[0], 0, "vwap_session_gap_was_bad must initialize to 0");
assert_eq!(c.last_bad_path[0], 0, "last_bad_path must initialize to 0 (no bad write)");
Ok(())
}