From eb9047fc30d90c2c13310163e1680ec03150ba72 Mon Sep 17 00:00:00 2001 From: jgrusewski Date: Fri, 15 May 2026 20:42:35 +0200 Subject: [PATCH] docs(phase-e): E.4 temporal encoder design + E.4.A implementation plan MIME-Version: 1.0 Content-Type: text/plain; charset=UTF-8 Content-Transfer-Encoding: 8bit Design doc (specs/): TFT-style architecture for Phase E execution policy — sliding window → Mamba2 SSM → GRN trunk → MoE regime gate → C51 head → Thompson selector. Two core pillars added per user: A) Full L1-L10 LOB depth input via hybrid MBP-10 peek B) ISV-continual-learning: controllers fire at training AND inference; Q-net weights frozen at inference but effective policy adapts via ISV modulation Plan doc (plans/): 14-task implementation plan for E.4.A foundation (window buffer + L1-L10 depth + Mamba2 forward+backward + ISV-eval controllers). Falsification gates: smoke R_mean improvement ≥ 50%, backtest cost=0 Sharpe ≥ +8 (no regression vs C51-flat +10.41), half-tick Sharpe ≥ -8 (closes 5pt+ of 10pt gap to Phase 1d.4 baseline -4.0). TGGN (foxhunt Temporal Graph Gated Network) explicitly deferred to Phase E.5+: existing CPU graph implementation + GPU adapter at ml-supervised/src/tgnn/ — marginal benefit for single-instrument ES futures vs the TFT-Mamba2 stack; revisit for multi-instrument extension or production HFT inference layer. Co-Authored-By: Claude Opus 4.7 --- ...6-05-15-phase-e-4-a-temporal-foundation.md | 1078 +++++++++++++++++ ...6-05-15-phase-e-temporal-encoder-design.md | 342 ++++++ 2 files changed, 1420 insertions(+) create mode 100644 docs/superpowers/plans/2026-05-15-phase-e-4-a-temporal-foundation.md create mode 100644 docs/superpowers/specs/2026-05-15-phase-e-temporal-encoder-design.md diff --git a/docs/superpowers/plans/2026-05-15-phase-e-4-a-temporal-foundation.md b/docs/superpowers/plans/2026-05-15-phase-e-4-a-temporal-foundation.md new file mode 100644 index 000000000..6fb1b3ca0 --- /dev/null +++ b/docs/superpowers/plans/2026-05-15-phase-e-4-a-temporal-foundation.md @@ -0,0 +1,1078 @@ +# Phase E.4.A — Temporal Foundation Implementation Plan + +> **For agentic workers:** REQUIRED SUB-SKILL: Use `superpowers:subagent-driven-development` (recommended) or `superpowers:executing-plans` to implement this plan task-by-task. Steps use checkbox (`- [ ]`) syntax for tracking. + +**Goal:** Lift the Phase E execution policy from a stateless C51 linear Q to a stateful Mamba2-encoded C51 Q, with real L1-L10 LOB depth wired through, and ISV controllers firing at inference. First A/B against the C51-flat baseline (cost=0 Sharpe_ann +10.41, half-tick -13.81). + +**Architecture:** state[10] window[K] + L1-L10 depth → VSN-deferred → **Mamba2 SSM** → **GRN trunk** → C51 head [9 actions × 51 atoms] → Thompson selector. ISV controllers fire at training AND backtest eval. Q-net weights frozen at inference; dynamic behavior via ISV modulation. + +**Tech Stack:** Rust 1.85+, CUDA 12.4, cudarc, existing foxhunt kernels (`mamba2_temporal_kernel.cu`, `grn_kernel.cu`, `alpha_c51.cu`) + new `alpha_window_push.cu`. SQLX_OFFLINE=true. + +**Predecessor design:** [2026-05-15-phase-e-temporal-encoder-design.md](../specs/2026-05-15-phase-e-temporal-encoder-design.md) + +**Falsification gate (must pass before E.4.B):** +- Smoke R_mean improvement ≥ 50% (≥ -0.5 vs C51-flat -1.1) +- rvr maintained ≥ +1.04σ +- Backtest cost=0 Sharpe_ann ≥ +8 (no regression vs C51-flat +10.41) +- Backtest half-tick Sharpe_ann ≥ -8 (5pt+ closure vs C51-flat -13.81) + +--- + +## File Structure + +**Will create:** +- `crates/ml/src/cuda_pipeline/alpha_window_push.cu` — circular-buffer push kernel +- `tests/integration/alpha_temporal_smoke_test.rs` — temporal foundation smoke test + +**Will modify:** +- `crates/ml/src/env/execution_env.rs` — extend `SnapshotRow.bid_l/ask_l` to `[f32; 10]` +- `crates/ml/src/env/loaders.rs` — hybrid MBP-10 peek in `load_snapshots_from_fxcache` +- `crates/ml/src/cuda_pipeline/alpha_kernels.rs` — wire Mamba2 / GRN launchers + new push kernel launcher +- `crates/ml/examples/alpha_dqn_h600_smoke.rs` — `--temporal` flag + Mamba2/GRN/C51 chain +- `crates/ml/examples/alpha_compose_backtest.rs` — `--temporal` flag + ISV controllers at eval +- `crates/ml/build.rs` — register `alpha_window_push.cu` + +**Will read (research only):** +- `crates/ml/src/cuda_pipeline/mamba2_temporal_kernel.cu` — extract kernel signature +- `crates/ml/src/cuda_pipeline/grn_kernel.cu` — extract kernel signature +- Production trainer integration sites (e.g., `crates/ml/src/trainers/dqn/trainer/training_loop.rs`) + +--- + +### Task 1: Research — Mamba2 + GRN integration map + +**Files (read-only):** +- Read: `crates/ml/src/cuda_pipeline/mamba2_temporal_kernel.cu` +- Read: `crates/ml/src/cuda_pipeline/grn_kernel.cu` +- Read: `crates/ml/src/trainers/dqn/trainer/training_loop.rs` (find Mamba2 + GRN call sites) + +**Goal:** produce a notebook entry `docs/superpowers/specs/2026-05-15-mamba2-grn-integration-notes.md` capturing: +- Mamba2 forward kernel function name + arg list + tensor shape contracts +- Mamba2 backward kernel function name + arg list +- Param-buffer layout (weight matrices A, B, C, Δ, x_proj, dt_bias, etc.) +- GRN forward kernel function name + arg list +- GRN backward kernel function name + arg list +- The production launcher signatures (if launchers exist in `crates/ml/src/cuda_pipeline/`) + +- [ ] **Step 1: Locate Mamba2 forward in source** + +```bash +grep -n "extern \"C\" __global__ void" crates/ml/src/cuda_pipeline/mamba2_temporal_kernel.cu | head -20 +``` + +Expected: list of kernel names (e.g., `mamba2_forward_kernel`, `mamba2_backward_kernel`). + +- [ ] **Step 2: Locate launcher calls in trainer** + +```bash +grep -rn "mamba2\|launch_mamba2\|GRN\|launch_grn" crates/ml/src/trainers/dqn/ crates/ml/src/cuda_pipeline/*.rs | head -30 +``` + +Identify: where Mamba2 forward is called from Rust, what host args are passed, what buffer sizes are expected. + +- [ ] **Step 3: Capture findings to notebook** + +Write `docs/superpowers/specs/2026-05-15-mamba2-grn-integration-notes.md` with sections: +- Mamba2 forward signature + buffer contract +- Mamba2 backward signature +- GRN forward + backward signatures +- Existing Rust launchers (if any) +- Integration risks identified + +- [ ] **Step 4: Commit notebook** + +```bash +git add docs/superpowers/specs/2026-05-15-mamba2-grn-integration-notes.md +git commit -m "docs(phase-e-4-a): mamba2 + grn integration research notes" +``` + +--- + +### Task 2: Research — production controller-at-inference pattern + +**Files (read-only):** +- Read: `crates/ml/src/trainers/dqn/trainer/training_loop.rs` for where controllers fire +- Read: `crates/ml/src/cuda_pipeline/alpha_kernels.rs:251-302` (existing `launch_stacker_threshold_controller`) +- Read: `crates/ml/src/cuda_pipeline/stacker_threshold_controller.cu` + +**Goal:** confirm the controller kernel is stateless across train/eval (it should be — reads/writes ISV + Wiener state buffer; not gated by training mode). Document any constraints that would prevent firing at eval. + +- [ ] **Step 1: Verify controller is mode-agnostic** + +```bash +grep -n "if.*training\|in_eval\|train_mode" crates/ml/src/cuda_pipeline/stacker_threshold_controller.cu +``` + +Expected: no branches on training-mode (controller does the same math always). + +- [ ] **Step 2: Identify Wiener state lifecycle** + +In `alpha_dqn_h600_smoke.rs`, find `ctl_wiener_dev` and note: how is it initialized, when is it reset. + +- [ ] **Step 3: Update integration notes** + +Append to `docs/superpowers/specs/2026-05-15-mamba2-grn-integration-notes.md`: +- Controller is mode-agnostic ✅ / requires modification +- Wiener state: how to handle at eval (preserve across cost cells, or reset per cell) + +- [ ] **Step 4: Commit** + +```bash +git add docs/superpowers/specs/2026-05-15-mamba2-grn-integration-notes.md +git commit -m "docs(phase-e-4-a): controller-at-inference verification" +``` + +--- + +### Task 3: Extend `SnapshotRow.bid_l/ask_l` to `[f32; 10]` + +**Files:** +- Modify: `crates/ml/src/env/execution_env.rs:73-86` (the `SnapshotRow` struct) +- Modify: `crates/ml/src/env/execution_env.rs:209-274` (env.step reads `bid_l[0]`, `ask_l[0]`, `bid_l[level]` for L1/L2; verify still works) +- Modify: `crates/ml/src/env/execution_env.rs:361-383` (synthetic_snapshots in tests) +- Modify: `crates/ml/src/env/loaders.rs:138-180` (synthetic L2/L3 lines must extend to L10) +- Modify: `crates/ml/examples/alpha_dqn_h600_smoke.rs:228-243` (MBP-10 mode synthesis) +- Modify: `crates/ml/examples/alpha_compose_backtest.rs:651-654` (the `let _: Option = None;` SnapshotRow ergonomics touchup) +- Test: `crates/ml/src/env/execution_env.rs` add unit tests at the bottom + +- [ ] **Step 1: Write the failing test** + +In `crates/ml/src/env/execution_env.rs` test module: + +```rust +#[test] +fn bid_l_and_ask_l_are_ten_deep() { + let row = SnapshotRow { + mid_price: 4500.0, + bid_l: [4499.75, 4499.5, 4499.25, 4499.0, 4498.75, + 4498.5, 4498.25, 4498.0, 4497.75, 4497.5], + ask_l: [4500.25, 4500.5, 4500.75, 4501.0, 4501.25, + 4501.5, 4501.75, 4502.0, 4502.25, 4502.5], + alpha_logit: 0.0, alpha_confidence: 0.0, + spread_bps: 5.0, l1_imbalance: 0.5, ofi_sum_5: 0.0, + mid_drift_5: 0.0, time_since_trade_s: 0.0, book_event_rate: 5.0, + }; + assert_eq!(row.bid_l.len(), 10); + assert_eq!(row.ask_l.len(), 10); + assert!(row.bid_l[0] > row.bid_l[9]); + assert!(row.ask_l[0] < row.ask_l[9]); +} +``` + +- [ ] **Step 2: Run test to verify it fails** + +```bash +SQLX_OFFLINE=true cargo test -p ml --lib env::execution_env::tests::bid_l_and_ask_l_are_ten_deep +``` + +Expected: compile error or assertion failure (current `bid_l: [f32; 3]`). + +- [ ] **Step 3: Edit `SnapshotRow` struct** + +Change in `execution_env.rs:73-86`: + +```rust +#[derive(Debug, Clone)] +pub struct SnapshotRow { + pub mid_price: f32, + pub bid_l: [f32; 10], + pub ask_l: [f32; 10], + pub alpha_logit: f32, + pub alpha_confidence: f32, + pub spread_bps: f32, + pub l1_imbalance: f32, + pub ofi_sum_5: f32, + pub mid_drift_5: f32, + pub time_since_trade_s: f32, + pub book_event_rate: f32, +} +``` + +- [ ] **Step 4: Update all SnapshotRow initializers** + +For each call site, extend the array literal: +- `loaders.rs:143-146` — replace synthesis with 10-level offsets: + +```rust +let mut bid_l = [0.0_f32; 10]; +let mut ask_l = [0.0_f32; 10]; +for k in 0..10 { + bid_l[k] = bid_l1 - (k as f32) * TICK; + ask_l[k] = ask_l1 + (k as f32) * TICK; +} +``` + +- `alpha_dqn_h600_smoke.rs:228-229` — same pattern +- `execution_env.rs:361-383` — `synthetic_snapshots` in tests, extend to 10 + +- [ ] **Step 5: Run all unit tests** + +```bash +SQLX_OFFLINE=true cargo test -p ml --lib env:: +``` + +Expected: PASS. If env.step tests fail, check that L1 (index 0) and L2 (index 1) usage is preserved. + +- [ ] **Step 6: Commit** + +```bash +git add crates/ml/src/env/ crates/ml/examples/alpha_dqn_h600_smoke.rs +git commit -m "feat(phase-e-4-a): extend SnapshotRow bid_l/ask_l to L1-L10" +``` + +--- + +### Task 4: Hybrid MBP-10 peek in fxcache loader + +**Files:** +- Modify: `crates/ml/src/env/loaders.rs:124-202` (add `mbp10_dir: Option<&Path>` arg) +- Modify: `crates/ml/examples/alpha_dqn_h600_smoke.rs:632-639` (caller) +- Modify: `crates/ml/examples/alpha_compose_backtest.rs:380-385` (caller) +- Test: integration in `crates/ml/src/env/loaders.rs` + +- [ ] **Step 1: Write the failing test** + +In `loaders.rs` add tests at bottom: + +```rust +#[test] +fn fxcache_loader_with_no_mbp10_synthesizes_depth() { + // Use a small fxcache fixture (if available) or skip with #[ignore] + // Confirms: when mbp10_dir is None, bid_l[3] = bid_l[0] - 3*TICK + // (synthesized as today's behavior) +} +``` + +Mark `#[ignore]` if no fixture; test is documentation-only. + +- [ ] **Step 2: Run test to verify it fails or runs ignored** + +```bash +SQLX_OFFLINE=true cargo test -p ml --lib env::loaders -- --ignored +``` + +- [ ] **Step 3: Add MBP-10 peek parameter** + +In `loaders.rs`, change the function signature: + +```rust +pub fn load_snapshots_from_fxcache( + fxcache_path: &Path, + max_snapshots: usize, + alpha_cache: Option<&[f32]>, + use_real_spread: bool, + mbp10_dir: Option<&Path>, // NEW +) -> Result> { +``` + +Inside the loop, after `let bid_l1 = ...; let ask_l1 = ...;`, add: + +```rust +let mut bid_l = [0.0_f32; 10]; +let mut ask_l = [0.0_f32; 10]; +bid_l[0] = bid_l1; +ask_l[0] = ask_l1; + +if let Some(_mbp10) = mbp10_dir { + // Phase E.4.A.2: real L4-L10 peek deferred to this task's follow-on + // commit — for the initial wiring, synthesize L2-L10 as before. + // The MBP-10 timestamp index requires fxcache bar-end-time access + // which isn't on the reader API yet. + for k in 1..10 { + bid_l[k] = bid_l1 - (k as f32) * TICK; + ask_l[k] = ask_l1 + (k as f32) * TICK; + } +} else { + for k in 1..10 { + bid_l[k] = bid_l1 - (k as f32) * TICK; + ask_l[k] = ask_l1 + (k as f32) * TICK; + } +} +``` + +**Note:** the FIRST commit wires the parameter only (synthesized output). Real MBP-10 peek is a follow-on commit. This split keeps each task small. The contract / signature change ships first. + +- [ ] **Step 4: Update callers** + +In `alpha_dqn_h600_smoke.rs:632-639` and `alpha_compose_backtest.rs:380-385`, append `None,` (or `cli.mbp10_dir.as_deref()` if a CLI arg exists; for now `None,`). + +- [ ] **Step 5: Run cargo check** + +```bash +SQLX_OFFLINE=true cargo check -p ml --release +``` + +Expected: clean. + +- [ ] **Step 6: Commit** + +```bash +git add crates/ml/src/env/loaders.rs crates/ml/examples/ +git commit -m "feat(phase-e-4-a): add mbp10_dir parameter to fxcache loader (synth-only impl)" +``` + +- [ ] **Step 7: Follow-on commit — real MBP-10 peek** + +This is a separate commit (next task focus). Skipped here so the parameter-introduction is isolated. + +--- + +### Task 5: Real MBP-10 peek implementation + +**Files:** +- Modify: `crates/ml/src/env/loaders.rs` (the `if let Some(mbp10) = mbp10_dir` branch) +- Read: `crates/ml-alpha/src/fxcache_reader.rs` (does the reader expose bar timestamp?) +- Read: `crates/data/src/providers/databento/dbn_parser.rs:704-728` (mbp10_streaming closure pattern) + +- [ ] **Step 1: Verify fxcache bar timestamps are accessible** + +```bash +grep -n "timestamp\|ts_event\|bar_time" crates/ml-alpha/src/fxcache_reader.rs | head -10 +``` + +If `reader.bar_timestamp(i)` or similar exists, proceed. If NOT, this task BLOCKS — fxcache reader needs an extension first (out-of-scope for E.4.A; document and skip the real-peek). + +- [ ] **Step 2: Write failing test** + +```rust +#[test] +#[ignore = "requires test fixtures: fxcache + matching mbp10 dir"] +fn fxcache_loader_with_mbp10_uses_real_l4_to_l10() { + let fxcache_path = std::path::Path::new("test_data/feature-cache/test.fxcache"); + let mbp10_dir = std::path::Path::new("test_data/futures-baseline-mbp10/ES.FUT"); + let rows = load_snapshots_from_fxcache( + fxcache_path, 100, None, false, Some(mbp10_dir), + ).unwrap(); + // The 4th level price gap from L1 should NOT be exactly 3*TICK + // (since real data has variable level spacings) + let first = &rows[0]; + let synth_bid_l3 = first.bid_l[0] - 3.0 * 0.25; + assert!((first.bid_l[3] - synth_bid_l3).abs() > 0.01, + "L4 bid should differ from synthesized offset"); +} +``` + +- [ ] **Step 3: Implement the timestamp-aligned MBP-10 peek** + +Inside `loaders.rs` real-peek branch: + +```rust +// Build a flat (timestamp, levels[10]) stream from MBP-10 .dbn[.zst] +// files in mbp10_dir, then for each fxcache bar i, find the last +// MBP-10 snapshot at-or-before bar_timestamp(i). +let parser = data::providers::databento::DbnParser::new() + .context("DbnParser for hybrid loader")?; +let files = crate::trainers::dqn::collect_dbn_files_recursive(mbp10); +let mut depth_stream: Vec<(i64, [f32; 10], [f32; 10])> = Vec::new(); +for file in &files { + let _ = parser.parse_mbp10_streaming(file, 1, |snap| { + if snap.levels.is_empty() { return; } + let mut bid = [0.0_f32; 10]; + let mut ask = [0.0_f32; 10]; + for k in 0..10.min(snap.levels.len()) { + bid[k] = raw_price_to_f32(snap.levels[k].bid_px); + ask[k] = raw_price_to_f32(snap.levels[k].ask_px); + } + depth_stream.push((snap.ts_event, bid, ask)); + }); +} +depth_stream.sort_by_key(|t| t.0); +// Now lookup per bar; use binary search on depth_stream. +// (Implementation continues...) +``` + +Note: this is sketch-level. The full implementation needs `raw_price_to_f32` helper and `collect_dbn_files_recursive` accessible (already exists in `ml::trainers::dqn`). + +- [ ] **Step 4: Wire `--use-real-depth` CLI flag in smoke binary** + +In `alpha_dqn_h600_smoke.rs`, add: + +```rust +/// Phase E.4.A: peek MBP-10 alongside fxcache to populate L4-L10 from +/// real depth data instead of synthesized ±tick offsets. Requires +/// `--mbp10-dir` to point at a directory with .dbn[.zst] files +/// covering the same time range as the fxcache. +#[arg(long, default_value_t = false)] +use_real_depth: bool, +``` + +In the loader call: + +```rust +let mbp10_for_peek = if cli.use_real_depth { + cli.mbp10_dir.as_deref() +} else { None }; +ml::env::loaders::load_snapshots_from_fxcache( + fxc, cli.max_snapshots, alpha_cache_vec.as_deref(), + cli.real_spread, mbp10_for_peek, +)? +``` + +Same in `alpha_compose_backtest.rs` (add `--mbp10-dir` + `--use-real-depth` flags; both default off). + +- [ ] **Step 5: cargo check** + +```bash +SQLX_OFFLINE=true cargo check -p ml --release +``` + +- [ ] **Step 6: Commit** + +```bash +git commit -am "feat(phase-e-4-a): real MBP-10 L4-L10 peek with --use-real-depth" +``` + +--- + +### Task 6: GPU window buffer + push kernel + +**Files:** +- Create: `crates/ml/src/cuda_pipeline/alpha_window_push.cu` +- Modify: `crates/ml/build.rs` (register the new cubin) +- Modify: `crates/ml/src/cuda_pipeline/alpha_kernels.rs` (add `ALPHA_WINDOW_PUSH_CUBIN` static + `launch_alpha_window_push`) +- Test: `crates/ml/src/cuda_pipeline/alpha_kernels.rs` (new test in test module) + +- [ ] **Step 1: Write the kernel** + +Create `crates/ml/src/cuda_pipeline/alpha_window_push.cu`: + +```c +// crates/ml/src/cuda_pipeline/alpha_window_push.cu +// +// Phase E.4.A.4 (2026-05-15): circular-buffer push kernel for the +// sliding-window state input to the Mamba2 temporal encoder. +// +// Single block, single thread copy: writes `state_dim` floats from +// `state_in[]` (mapped-pinned or device pointer) into +// `window[head_idx * state_dim .. (head_idx+1) * state_dim]`. +// +// The host increments `head_idx = (head_idx + 1) % K` after each call. +// On episode reset the host zeros the window buffer via a separate +// `cudaMemsetAsync`. + +#include + +extern "C" __global__ void alpha_window_push_kernel( + const float* __restrict__ state_in, // [state_dim] + float* __restrict__ window, // [K, state_dim] row-major + int head_idx, + int state_dim +) { + const int j = blockIdx.x * blockDim.x + threadIdx.x; + if (j >= state_dim) return; + window[head_idx * state_dim + j] = state_in[j]; +} +``` + +- [ ] **Step 2: Register in build.rs** + +In `crates/ml/build.rs`, find the list `kernels_with_common` and add after `"alpha_c51.cu"`: + +```rust + // Phase E.4.A.4 (2026-05-15): circular-buffer push for sliding- + // window state input to Mamba2 temporal encoder. + "alpha_window_push.cu", +``` + +- [ ] **Step 3: Add launcher + cubin static in alpha_kernels.rs** + +```rust +/// Precompiled circular-buffer window push cubin. Phase E.4.A. +pub static ALPHA_WINDOW_PUSH_CUBIN: &[u8] = + include_bytes!(concat!(env!("OUT_DIR"), "/alpha_window_push.cubin")); + +/// Launch `alpha_window_push_kernel`. Copies the current state[state_dim] +/// into the circular buffer slot indexed by `head_idx`. +/// +/// # Safety +/// `state_in_dev` and `window_dev` MUST be valid device pointers. +/// `window_dev` capacity ≥ `K * state_dim` floats. +pub unsafe fn launch_alpha_window_push( + stream: &cudarc::driver::CudaStream, + kernel: &cudarc::driver::CudaFunction, + state_in_dev: u64, + window_dev: u64, + head_idx: i32, + state_dim: i32, +) -> Result<(), MLError> { + use cudarc::driver::{LaunchConfig, PushKernelArg}; + + debug_assert!(state_dim > 0, "state_dim must be positive"); + debug_assert!(head_idx >= 0, "head_idx must be non-negative"); + + const BLOCK: u32 = 32; + let grid_x = ((state_dim as u32) + BLOCK - 1) / BLOCK; + let cfg = LaunchConfig { + grid_dim: (grid_x.max(1), 1, 1), + block_dim: (BLOCK, 1, 1), + shared_mem_bytes: 0, + }; + stream + .launch_builder(kernel) + .arg(&state_in_dev) + .arg(&window_dev) + .arg(&head_idx) + .arg(&state_dim) + .launch(cfg) + .map_err(|e| MLError::ModelError(format!("alpha_window_push launch: {e}")))?; + Ok(()) +} +``` + +- [ ] **Step 4: Write the GPU unit test** + +In `alpha_kernels.rs` test module, add: + +```rust +#[test] +fn alpha_window_push_circular_writes() -> anyhow::Result<()> { + let ctx = CudaContext::new(0)?; + let stream = ctx.default_stream(); + let module = ctx.load_cubin(ALPHA_WINDOW_PUSH_CUBIN.to_vec())?; + let kernel = module.load_function("alpha_window_push_kernel")?; + + const K: usize = 4; + const D: usize = 3; + let mut window_dev = stream.alloc_zeros::(K * D)?; + let state_a: Vec = vec![1.0, 2.0, 3.0]; + let state_b: Vec = vec![10.0, 20.0, 30.0]; + let state_dev = stream.clone_htod(&state_a)?; + + // Push state_a at head_idx=0 + { + let (s_ptr, _g0) = state_dev.device_ptr(&stream); + let (w_ptr, _g1) = window_dev.device_ptr_mut(&stream); + unsafe { + launch_alpha_window_push(&stream, &kernel, s_ptr, w_ptr, 0, D as i32)?; + } + } + // Push state_b at head_idx=2 + let state_dev_b = stream.clone_htod(&state_b)?; + { + let (s_ptr, _g0) = state_dev_b.device_ptr(&stream); + let (w_ptr, _g1) = window_dev.device_ptr_mut(&stream); + unsafe { + launch_alpha_window_push(&stream, &kernel, s_ptr, w_ptr, 2, D as i32)?; + } + } + let window_host = stream.clone_dtoh(&window_dev)?; + assert_eq!(&window_host[0..3], &[1.0, 2.0, 3.0]); + assert_eq!(&window_host[3..6], &[0.0, 0.0, 0.0]); + assert_eq!(&window_host[6..9], &[10.0, 20.0, 30.0]); + assert_eq!(&window_host[9..12], &[0.0, 0.0, 0.0]); + Ok(()) +} +``` + +- [ ] **Step 5: Run test** + +```bash +SQLX_OFFLINE=true cargo test -p ml --release --lib cuda_pipeline::alpha_kernels::tests::alpha_window_push_circular_writes +``` + +Expected: PASS. + +- [ ] **Step 6: Commit** + +```bash +git add crates/ml/src/cuda_pipeline/alpha_window_push.cu crates/ml/build.rs crates/ml/src/cuda_pipeline/alpha_kernels.rs +git commit -m "feat(phase-e-4-a): alpha_window_push circular-buffer kernel" +``` + +--- + +### Task 7: Wire window buffer into smoke binary (training) + +**Files:** +- Modify: `crates/ml/examples/alpha_dqn_h600_smoke.rs` — add window buffer + push kernel call per step + episode-reset clearing + +- [ ] **Step 1: Add `--temporal` CLI flag + `--window-k`** + +```rust +/// Phase E.4.A: enable the temporal-encoder stack (sliding window → +/// Mamba2 → GRN → C51). When false, runs the existing C51-flat path. +#[arg(long, default_value_t = false)] +temporal: bool, +/// Sliding-window length K. 16 starting point; sweep 16/32/64. +#[arg(long, default_value_t = 16)] +window_k: usize, +``` + +- [ ] **Step 2: Load window-push kernel + allocate window buffer** + +After other cubin loads, add: + +```rust +let push_module = ctx.load_cubin(ml::cuda_pipeline::alpha_kernels::ALPHA_WINDOW_PUSH_CUBIN.to_vec())?; +let push_kernel = push_module.load_function("alpha_window_push_kernel")?; +``` + +After other device allocations, add: + +```rust +let mut window_dev = stream.alloc_zeros::(cli.window_k * STATE_DIM)?; +let mut head_idx: i32 = 0; +``` + +- [ ] **Step 3: Push per step + advance head_idx** + +Inside the per-step loop, REPLACE the existing single-state upload with: + +```rust +state_pinned.write(&s_vec); // already exists for C51 path +if cli.temporal { + let (s_ptr, _g0) = ...; // state_pinned.dev_u64() + let (w_ptr, _g1) = window_dev.device_ptr_mut(&stream); + unsafe { + ml::cuda_pipeline::alpha_kernels::launch_alpha_window_push( + &stream, &push_kernel, + state_pinned.dev_u64(), w_ptr, + head_idx, STATE_DIM as i32, + )?; + } + head_idx = (head_idx + 1) % (cli.window_k as i32); +} +``` + +- [ ] **Step 4: Zero window + reset head_idx on episode start** + +Before the inner `loop {` in the training loop: + +```rust +if cli.temporal { + stream.memset_zeros(&mut window_dev)?; + head_idx = 0; +} +``` + +(or equivalent `stream.alloc_zeros` reuse — check cudarc API for memset). + +- [ ] **Step 5: cargo check** + +```bash +SQLX_OFFLINE=true cargo check -p ml --release --example alpha_dqn_h600_smoke +``` + +Expected: clean. + +- [ ] **Step 6: Run smoke with --temporal (window-only; no Mamba2 yet)** + +The forward path STILL uses the C51 forward on state_pinned (not h_s2). This commit just adds the buffer maintenance. We expect bit-identical behavior to the existing C51 path — the buffer is populated but not consumed by any kernel yet. + +```bash +SQLX_OFFLINE=true cargo run -p ml --release --example alpha_dqn_h600_smoke -- \ + --fxcache-path /home/jgrusewski/Work/foxhunt/test_data/feature-cache/9297017b6db6795f75e57be8aefb03e45e6427513f42c08e22521e58fa025d5e.fxcache \ + --alpha-cache config/ml/alpha_logits_cache.bin \ + --fill-coeffs config/ml/alpha_fill_coeffs.json \ + --horizon 600 --n-episodes 100 --c51 --temporal --window-k 16 \ + --out-path /tmp/alpha_dqn_h600_smoke_temporal_buffer_only.json +``` + +Expected: runs to completion. R_mean trajectory matches the C51-flat run within noise. + +- [ ] **Step 7: Commit** + +```bash +git commit -am "feat(phase-e-4-a): sliding-window buffer + push call site in smoke" +``` + +--- + +### Task 8: Wire Mamba2 forward in smoke binary + +**Files (depends on Task 1's research notes):** +- Modify: `crates/ml/examples/alpha_dqn_h600_smoke.rs` — Mamba2 forward chain +- Read: `docs/superpowers/specs/2026-05-15-mamba2-grn-integration-notes.md` (Task 1's output for signature) +- Modify: `crates/ml/src/cuda_pipeline/alpha_kernels.rs` — if Mamba2 launcher doesn't exist publicly, add a `launch_mamba2_forward_alpha` wrapper + +- [ ] **Step 1: Load the Mamba2 cubin + kernel function** + +Using the function name from Task 1's notes (placeholder: `mamba2_forward_kernel`). + +```rust +let mamba2_module = ctx.load_cubin(ml::cuda_pipeline::alpha_kernels::ALPHA_MAMBA2_CUBIN.to_vec())?; +let mamba2_fwd_kernel = mamba2_module.load_function("mamba2_forward_kernel")?; +``` + +(If `ALPHA_MAMBA2_CUBIN` doesn't exist, use `MAMBA2_TEMPORAL_CUBIN` from production or include_bytes! directly from OUT_DIR.) + +- [ ] **Step 2: Allocate Mamba2 parameter buffers** + +Per the kernel signature (Task 1 notes), allocate: +- A, B, C matrices (SSM core) +- Δ (delta projection) +- x_proj +- dt_bias +- output buffer `h_temporal` of size [B, H] where H is the Mamba2 hidden dim + +```rust +const MAMBA2_HIDDEN_DIM: usize = 32; // start small; tune later +let mut mamba2_params: ...; +let mut h_temporal_dev = stream.alloc_zeros::(MAMBA2_HIDDEN_DIM)?; +``` + +(Detailed sizing depends on Task 1 notes.) + +- [ ] **Step 3: Replace the forward call** + +In the inference path's `if cli.temporal { ... }` branch: + +```rust +// Run Mamba2 forward on the window buffer +unsafe { + launch_mamba2_forward( + &stream, &mamba2_fwd_kernel, + window_dev_ptr, // [K, state_dim] + mamba2_params, + h_temporal_dev_ptr, // [H] + cli.window_k as i32, + STATE_DIM as i32, + MAMBA2_HIDDEN_DIM as i32, + )?; +} +// Then C51 forward on h_temporal instead of state_pinned +``` + +C51 forward must take `h_temporal` (size H = 32) instead of state (size 10). This means changing the C51 W layout from `[9·51, 10]` to `[9·51, 32]`. + +- [ ] **Step 4: Adapt C51 Q-net dimensions** + +```rust +let c51_input_dim = if cli.temporal { MAMBA2_HIDDEN_DIM } else { STATE_DIM }; +let n_weights_eff = N_ACTIONS * c51_n_atoms * c51_input_dim; +// w_init.len() = n_weights_eff; existing code uses STATE_DIM in xavier init — update. +``` + +In the C51 forward call, pass `c51_input_dim as i32` instead of `state_dim_i`. + +- [ ] **Step 5: cargo check** + +```bash +SQLX_OFFLINE=true cargo check -p ml --release --example alpha_dqn_h600_smoke +``` + +- [ ] **Step 6: Run minimal smoke** + +```bash +SQLX_OFFLINE=true cargo run -p ml --release --example alpha_dqn_h600_smoke -- \ + --fxcache-path ... --alpha-cache ... --fill-coeffs ... \ + --horizon 600 --n-episodes 100 --c51 --temporal --window-k 16 \ + --out-path /tmp/alpha_dqn_smoke_mamba2_only.json +``` + +Expected: runs to completion. R_mean trajectory may differ from C51-flat (different architecture). + +- [ ] **Step 7: Commit** + +```bash +git commit -am "feat(phase-e-4-a): wire Mamba2 forward in smoke --temporal path" +``` + +--- + +### Task 9: Wire GRN trunk + +**Files:** +- Modify: `crates/ml/examples/alpha_dqn_h600_smoke.rs` — add GRN forward after Mamba2 + +- [ ] **Step 1: Load GRN cubin + kernel** + +```rust +let grn_module = ctx.load_cubin(...)?; +let grn_fwd_kernel = grn_module.load_function("grn_forward_kernel")?; +``` + +(Exact name from Task 1 notes.) + +- [ ] **Step 2: Allocate GRN params + output** + +Per Task 1 sizing. GRN typical: H_in → H' → H_in with residual gate. Output `h_s2` same size as Mamba2's H or projected. + +- [ ] **Step 3: Insert GRN between Mamba2 and C51** + +```rust +// After Mamba2 forward: +launch_grn_forward(... mamba2_h_temporal_dev, grn_params, h_s2_dev ...)?; +// Then C51 reads h_s2 instead of h_temporal +``` + +- [ ] **Step 4: cargo check + smoke run + commit** + +Same as Task 8. + +```bash +git commit -am "feat(phase-e-4-a): wire GRN trunk in smoke --temporal" +``` + +--- + +### Task 10: Wire Mamba2 + GRN backward + +**Files:** +- Modify: `crates/ml/examples/alpha_dqn_h600_smoke.rs` — add backward calls before SGD step +- Modify: `crates/ml/src/cuda_pipeline/alpha_kernels.rs` — if launchers don't exist publicly, add wrappers + +- [ ] **Step 1: Identify backward kernel signatures** + +From Task 1 notes: `grn_backward_kernel`, `mamba2_backward_kernel`. Note backward output buffers (dW, db, dx_in). + +- [ ] **Step 2: Allocate gradient buffers for Mamba2 + GRN params** + +Same shapes as the param buffers from Tasks 8/9. + +- [ ] **Step 3: Insert backward calls between C51 grad and SGD step** + +```rust +// C51 grad already produces dW_c51, db_c51 + dh_s2 (gradient into GRN output) +// Run GRN backward: dh_s2 → dgrn_params, dh_temporal +// Run Mamba2 backward: dh_temporal → dmamba2_params, dwindow_buffer +// Then SGD step on Mamba2 + GRN + C51 params +``` + +- [ ] **Step 4: cargo check + smoke run + commit** + +```bash +git commit -am "feat(phase-e-4-a): Mamba2 + GRN backward in smoke --temporal" +``` + +--- + +### Task 11: ISV controllers fire at backtest eval + +**Files:** +- Modify: `crates/ml/examples/alpha_compose_backtest.rs` — add controller invocation per eval episode + +- [ ] **Step 1: Add `--isv-continual` CLI flag** + +```rust +/// Phase E.4.A.7: ISV-continual-learning. When true, the +/// stacker-threshold + Kelly-attenuation controller fires at the end +/// of EVERY eval episode (not just training). Slot 543 (threshold) +/// adapts during the 2D sweep. +#[arg(long, default_value_t = true)] +isv_continual: bool, +``` + +- [ ] **Step 2: Load controller kernel + Wiener state at backtest start** + +Mirror the smoke binary's setup: load `STACKER_THRESHOLD_CONTROLLER_CUBIN`, allocate `ctl_wiener_dev`, set `isv_host[TRADE_RATE_TARGET_INDEX]`. + +- [ ] **Step 3: Inside the eval episode loop, fire the controller** + +After each episode's terminal reward is observed: + +```rust +if cli.isv_continual { + let trade_count_ep = ep_n_trades as f32; + let decisions_ep = (state.step as f32).max(1.0); + let baseline_std = isv_host[RANDOM_BASELINE_STD_INDEX].max(1e-6); + let rollout_sharpe = terminal_r / baseline_std; + unsafe { + let (isv_ptr, _g0) = isv_dev.device_ptr_mut(&stream); + let (w_ptr, _g1) = ctl_wiener_dev.device_ptr_mut(&stream); + ml::cuda_pipeline::alpha_kernels::launch_stacker_threshold_controller( + &stream, &ctl_kernel, + trade_count_ep, decisions_ep, rollout_sharpe, + /* target_sharpe */ 0.5, + /* k_threshold */ 0.01, + /* k_atten */ 0.005, + /* wiener_alpha_floor */ 0.4, + /* ctl_alpha_meta */ 0.1, + STACKER_THRESHOLD_INDEX as i32, + TRADE_RATE_TARGET_INDEX as i32, + TRADE_RATE_OBSERVED_EMA_INDEX as i32, + STACKER_KELLY_ATTENUATION_INDEX as i32, + isv_ptr, w_ptr, + )?; + } + // Refresh CPU threshold cache for next episode's gating + stream.synchronize()?; + let isv_now = stream.clone_dtoh(&isv_dev)?; + let current_threshold = isv_now[STACKER_THRESHOLD_INDEX]; + // The eval `threshold` param used in epsilon_greedy_gated is the + // SWEEP threshold (fixed per cell). Decide: does ISV-continual + // OVERRIDE the sweep grid, or compose with it? For the first cut, + // use ISV threshold as a floor — `actual_thresh = max(threshold, current_threshold)`. +} +``` + +(Decision: ISV-continual at backtest is a parallel observation, NOT replacing the τ-grid. Document in code why.) + +- [ ] **Step 4: cargo check + run backtest** + +```bash +SQLX_OFFLINE=true cargo run -p ml --release --example alpha_compose_backtest -- \ + --fxcache-path ... --alpha-cache ... --fill-coeffs ... \ + --c51 --isv-continual \ + --out-path /tmp/alpha_compose_backtest_c51_isv_continual.json +``` + +Compare best-per-cost Sharpe_ann to the C51 baseline (+10.41 / -13.81). + +- [ ] **Step 5: Commit** + +```bash +git commit -am "feat(phase-e-4-a): ISV controllers fire at backtest eval" +``` + +--- + +### Task 12: Mirror --temporal path in backtest binary + +**Files:** +- Modify: `crates/ml/examples/alpha_compose_backtest.rs` — copy the smoke binary's temporal forward/backward chain + +- [ ] **Step 1: Add CLI flags** + +`--temporal`, `--window-k`, `--mbp10-dir`, `--use-real-depth`. + +- [ ] **Step 2: Load Mamba2 + GRN + window-push kernels** + +Mirror smoke setup. + +- [ ] **Step 3: Update training inference + batched compute + eval inference** + +Three places to add `if cli.temporal { mamba2 → grn → c51 } else { c51-flat }` branches. + +- [ ] **Step 4: cargo check** + +```bash +SQLX_OFFLINE=true cargo check -p ml --release --example alpha_compose_backtest +``` + +- [ ] **Step 5: Commit** + +```bash +git commit -am "feat(phase-e-4-a): --temporal path in backtest binary" +``` + +--- + +### Task 13: Smoke validation + +**Files:** (no source changes; experiment + memory) + +- [ ] **Step 1: Run baseline C51-flat smoke** + +(Reference numbers exist from earlier; re-run if seed/data changed.) + +```bash +SQLX_OFFLINE=true cargo run -p ml --release --example alpha_dqn_h600_smoke -- \ + --fxcache-path /home/jgrusewski/Work/foxhunt/test_data/feature-cache/9297017b6db6795f75e57be8aefb03e45e6427513f42c08e22521e58fa025d5e.fxcache \ + --alpha-cache config/ml/alpha_logits_cache.bin \ + --fill-coeffs config/ml/alpha_fill_coeffs.json \ + --horizon 600 --n-episodes 1000 --c51 \ + --out-path /tmp/alpha_dqn_smoke_c51_flat_baseline.json +``` + +- [ ] **Step 2: Run temporal smoke** + +```bash +SQLX_OFFLINE=true cargo run -p ml --release --example alpha_dqn_h600_smoke -- \ + --fxcache-path ... --alpha-cache ... --fill-coeffs ... \ + --horizon 600 --n-episodes 1000 --c51 --temporal --window-k 16 \ + --out-path /tmp/alpha_dqn_smoke_temporal.json +``` + +- [ ] **Step 3: Compare R_mean trajectories** + +Read both JSONs, plot R_mean vs episode. Expected (per falsification gate): +- Temporal R_mean at ep 1000 ≥ -0.5 (vs C51-flat -1.1) — at least 50% improvement +- rvr ≥ +1.04 + +- [ ] **Step 4: Write memory file** + +Create `pearl_phase_e_4_a_temporal_smoke_verdict.md` capturing the comparison. + +- [ ] **Step 5: Commit memory + sweep JSONs** + +```bash +git add /home/jgrusewski/.claude/projects/-home-jgrusewski-Work-foxhunt/memory/pearl_phase_e_4_a_temporal_smoke_verdict.md +git commit -m "docs(phase-e-4-a): smoke verdict — temporal vs C51-flat" +``` + +--- + +### Task 14: Backtest validation (final E.4.A gate) + +**Files:** (no source changes) + +- [ ] **Step 1: Run baseline C51-flat backtest** + +```bash +SQLX_OFFLINE=true cargo run -p ml --release --example alpha_compose_backtest -- \ + --fxcache-path ... --alpha-cache ... --fill-coeffs ... --c51 \ + --out-path /tmp/alpha_compose_backtest_c51_flat.json +``` + +- [ ] **Step 2: Run temporal backtest with ISV-continual** + +```bash +SQLX_OFFLINE=true cargo run -p ml --release --example alpha_compose_backtest -- \ + --fxcache-path ... --alpha-cache ... --fill-coeffs ... \ + --c51 --temporal --window-k 16 --isv-continual \ + --out-path /tmp/alpha_compose_backtest_temporal_isv.json +``` + +- [ ] **Step 3: Compare per-cost best Sharpe_ann** + +Read both JSONs, build comparison table. Falsification gate: +- cost=0 Sharpe_ann ≥ +8 (vs C51-flat +10.41) — NO REGRESSION +- half-tick Sharpe_ann ≥ -8 (vs C51-flat -13.81) — closes 5pt+ of gap + +- [ ] **Step 4: Decision** + +- **PASS:** advance to Phase E.4.B (VSN + MoE regime gating). +- **PARTIAL (smoke pass, backtest below half-tick threshold):** diagnose; tune window K (try 32, 64); revalidate. +- **FAIL:** ROLLBACK. The temporal architecture isn't lifting on Phase E ES futures. Investigate before adding more E.4 features. + +- [ ] **Step 5: Write E.4.A close-out memory** + +Create `project_phase_e_4_a_close.md` with full backtest table + decision. + +- [ ] **Step 6: Update MEMORY.md index** + +Add the new memory file's one-line entry. + +- [ ] **Step 7: Final commit** + +```bash +git add /home/jgrusewski/.claude/projects/-home-jgrusewski-Work-foxhunt/memory/ +git commit -m "docs(phase-e-4-a): E.4.A close-out verdict + memory index" +``` + +--- + +## Self-Review + +**Spec coverage check (vs Section 6 of the design doc):** +- [x] Phase E.4.A.1 Sliding-window state buffer — Tasks 6-7 +- [x] Phase E.4.A.2 Hybrid MBP-10 depth loader — Tasks 4-5 +- [x] Phase E.4.A.3 Extend state vector for depth features — Task 3 (extends `bid_l/ask_l` to 10; depth-derived FEATURES in state vector deferred to Task 7 follow-on; flagged in note below) +- [x] Phase E.4.A.4 Mamba2 forward + backward — Tasks 8 + 10 +- [x] Phase E.4.A.5 GRN trunk — Tasks 9 + 10 backward +- [x] Phase E.4.A.6 C51 head reads h_s2 — Task 8 step 4 +- [x] Phase E.4.A.7 ISV-continual at eval — Task 11 + +**Known scope gap:** the design doc Phase E.4.A.3 asked for "depth-derived state features" (cumulative bid/ask size, spread curve slope). This plan defers those to a Phase E.4.A follow-on commit because they require either (a) extending SnapshotRow with size arrays AND mid-episode size-feature derivation, which is its own task scope. For the FIRST E.4.A validation, the state vector stays at 10-dim. If the temporal architecture itself lifts Sharpe (Task 14), depth features become a clear next iteration. + +**Type consistency check:** +- `STATE_DIM = 10` referenced consistently in Tasks 6-11 +- `MAMBA2_HIDDEN_DIM = 32` introduced in Task 8 and used in Tasks 9-10 +- `c51_input_dim` toggles between `STATE_DIM` and `MAMBA2_HIDDEN_DIM` based on `cli.temporal` +- Mamba2 + GRN kernel arg signatures are placeholders pending Task 1 research outputs — implementations must reconcile against the actual signatures discovered there + +**Placeholder check:** +- Several signatures in Tasks 8-10 are sketched against assumed kernel APIs. The Task 1 research notebook is the source-of-truth for actual signatures. Implementers MUST cross-check. + +--- + +## Execution Handoff + +Plan complete. Saved to `docs/superpowers/plans/2026-05-15-phase-e-4-a-temporal-foundation.md`. Two execution options: + +**1. Subagent-Driven (recommended)** — dispatch a fresh subagent per task, review between tasks, fast iteration. Uses `superpowers:subagent-driven-development`. + +**2. Inline Execution** — execute tasks in this session using `superpowers:executing-plans`. Batch execution with checkpoints for review. + +**Which approach?** diff --git a/docs/superpowers/specs/2026-05-15-phase-e-temporal-encoder-design.md b/docs/superpowers/specs/2026-05-15-phase-e-temporal-encoder-design.md new file mode 100644 index 000000000..0f13ec529 --- /dev/null +++ b/docs/superpowers/specs/2026-05-15-phase-e-temporal-encoder-design.md @@ -0,0 +1,342 @@ +# Phase E Temporal Encoder + Reasoning Architecture — Design + +**Status:** Design doc (not yet implementation plan). Awaiting user approval. +**Date:** 2026-05-15 +**Author:** session 9c4e48ad-0f44-4989-8dab-09ec6870540c +**Predecessors:** `pearl_c51_thompson_closed_phase_e3_gap.md`, `pearl_action_pruning_falsified.md`, `project_phase_e3_close.md` +**Production reference stack:** `mamba2_temporal_kernel.cu`, `tlob_kernel.cu`, `attention_kernel.cu`, `grn_kernel.cu`, `vsn_feature_selection_kernel.cu`, `moe_kernels.cu`, `aux_heads_kernel.cu`, `aux_trunk_forward_kernel.cu`, `pearl_1_atom_kernel.cu` + +## 1. Goal + +Lift the Phase E execution policy from **stateless linear C51 Q** to a **stateful, regime-aware, auxiliary-supervised** Q-network. Close the remaining 10pt Sharpe gap to the Phase 1d.4 baseline at half-tick by adopting the production trainer's temporal-reasoning stack — same components that lifted Phase 1d.2's snapshot-stream AUC from 0.50 → 0.66 at K=6000. + +## 2. Why now + +The Phase E.3 close + C51 follow-up establish: + +- ✅ Alpha signal is real and gets transmitted (rvr = +1.045σ across linear and C51). +- ✅ Calibration was the dominant bottleneck (C51 closed +26pt of Sharpe gap at cost=0 vs linear Q). +- ✅ Action variance was NOT the bottleneck (pruning falsified, -22pt Sharpe). +- ✅ Fill economics was NOT the bottleneck (real spread ≈ fixed for ES — 76% of bars at 1-tick floor). +- ⚠️ The remaining 10pt half-tick gap is **trade-count economics × value-estimation precision** under current stateless architecture. + +The production trainer already implements the techniques needed: +- **Temporal memory** (Mamba2, TLOB, attention) — `pearl_state_amplifies_short_horizon_into_long_horizon` proved this works on ES futures. +- **Regime-aware decisions** (MoE) — `pearl_snapshot_alpha_is_regime_conditional` showed spread-Q4 acc=0.747 vs middle quintiles below chance; the alpha lives in specific regimes. +- **Dense auxiliary supervision** (aux heads + separate aux trunk) — `pearl_separate_aux_trunk_when_shared_starves` is the canonical fix for sparse-PnL training. +- **Adaptive atom support** (`pearl_per_branch_c51_atom_span`) — eliminates the hardcoded [-10, +10] choice. + +Smart-borrow philosophy: adopt the **techniques** (kernels, controller signals, ISV slots) without importing the **specialization** (4-branch action factorization, magnitude bins, direction-specific reward biases). Same approach that worked for the C51 borrow. + +## 2.5. Core architectural pillars (added by user 2026-05-15) + +### Pillar A: Full L1-L10 LOB depth input + +The DBN MBP-10 files contain real L1-L10 bid/ask data (verified +`dbn_parser.rs:721-728` and `866-873` correctly copy all 10 levels). +The current Phase E env synthesizes L2/L3 at ±tick offsets because +the **fxcache** doesn't carry depth — only L1-derived features +(spread_bps, l1_imbalance) plus the 81-dim Block-S feature vector. + +To use real L4-L10, options: +1. **Hybrid loader**: at env construction, peek MBP-10 by timestamp + for each fxcache bar — adds depth without breaking alpha_cache + alignment. +2. **Fxcache rebuild**: extend the fxcache schema to store full depth. +3. **MBP-10 direct mode**: skip fxcache entirely; loses alpha_cache. + +Pick (1) for the staged rollout. Adds ~120 floats per snapshot +(L1-L10 × bid+ask × {px, sz, ct} = 60 fields × 2 sides = up to 120). +Phase E.4.A.2 task. + +### Pillar B: ISV-continual-learning (train AND inference) + +Foxhunt's controller pattern already produces and consumes ISV slots +GPU-resident via `pearl_engagement_rate_self_correction` and friends. +Phase E currently freezes ISV at training end. **The proposal: keep +controllers firing during inference / eval / deployment.** + +**What stays frozen at inference:** Q-net weights (W, b for the C51 +head; Mamba2 SSM parameters; GRN/MoE/VSN weights). + +**What adapts at inference via ISV controllers:** +- Slot 543 — stacker threshold (engagement-rate controller continues + responding to observed trade rate) +- Slot 545 — observed-rate Wiener-α EMA (always updating) +- Slot 546 — Kelly attenuation (tightens after drawdowns) +- Slot atom_headroom (Pearl-1) — adaptive C51 support widens/narrows + as observed Q-scale shifts +- Slot ~126 — MoE gate entropy EMA (drives MoE λ controller) +- ISV[12] — health composition (if we adopt it) gates ALL downstream + adaptations + +**Effective dynamic weight:** the policy's decision rule +`action = ThompsonSelect(MoE_mix(experts), threshold_gate)` is a +*function* of the ISV slots. As ISV evolves at inference, +the effective policy adapts WITHOUT touching neural-network parameters. +This is "dynamic weight via ISV modulation" — lightweight, safe, +composable with frozen Q-net. + +**Optional aggressive variant — LoRA at inference:** add a low-rank +adapter `W' = W + α·U·V^T` where `U ∈ R^{H×r}`, `V ∈ R^{r×H}`, +`r << H`. Update U, V via online SGD on observed reward error during +inference. Adds literal dynamic weights (not just ISV). Out of scope +for Phase E.4 core — possible Phase E.5 extension if ISV-only +adaptation isn't enough. + +**Why this composes:** Mamba2 hidden state accumulates within-episode +temporal context (fast adaptation). ISV controllers accumulate +across-episode regime context (slow adaptation). Together they form +a two-timescale online adaptation system on top of a frozen +representational core. + +**Failure modes to guard against:** +- ISV runaway in low-data regimes — bound all controller updates by + Wiener-α floor per `pearl_wiener_alpha_floor_for_nonstationary`. +- Threshold spiraling to no-trade — engagement-rate controller has + a target floor; verify it activates at inference. +- Adversarial market manipulation against the live ISV — for HFT + deployment, controllers should have rate limits on update magnitude. + +## 3. Architecture (target) + +``` +┌──────────────────────────────────────────────────────────────────────┐ +│ Per-step input: state[10] + rolling window buffer[K, 10] │ +│ (current state pushed into a circular buffer of last K snapshots) │ +└────────────────────────────────┬─────────────────────────────────────┘ + │ + ▼ +┌──────────────────────────────────────────────────────────────────────┐ +│ Variable Selection Network (VSN) │ +│ • Per-state softmax gate over feature groups → vsn_mask[10] │ +│ • Element-wise gated features → x_in[K, 10] │ +│ • mask EMA → ISV[~120..125] (interpretability) │ +└────────────────────────────────┬─────────────────────────────────────┘ + │ + ▼ +┌──────────────────────────────────────────────────────────────────────┐ +│ Mamba2 SSM Temporal Encoder │ +│ • State-space accumulation over the K-bar window │ +│ • Output: h_temporal ∈ R^H (compressed temporal context) │ +│ • SAME kernel as Phase 1d.2 — proven on ES futures │ +└────────────────────────────────┬─────────────────────────────────────┘ + │ + ▼ +┌──────────────────────────────────────────────────────────────────────┐ +│ Gated Residual Network (GRN) Trunk │ +│ • h_temporal → h_s2 ∈ R^H' (shared trunk encoding) │ +│ • Gating + residual; more expressive than linear, cheaper than │ +│ full transformer │ +└──────────────────────────────┬─┴─────────────────────────────────────┐ + │ │ + ┌────────────────┘ │ + │ (main path) │ + ▼ ▼ +┌──────────────────────────────────┐ ┌──────────────────────────────┐ +│ MoE Regime Gate │ │ Aux Trunk (SEPARATE) │ +│ • K_e=4 experts, gate by h_s2 │ │ • Linear→ELU→Linear→ELU→ │ +│ • Each expert is a C51 head │ │ Linear MLP │ +│ [9 actions × N_atoms atoms] │ │ • stop_grad at encoder │ +│ • Output: gate-mixed probs │ │ boundary │ +│ p(a, k | s) │ │ • Independent Adam │ +│ • gate_entropy_ema → ISV[126] │ │ • Aux heads: │ +│ drives MoE λ controller │ │ 1) next-bar return MSE │ +│ (existing in production) │ │ 2) 5-class regime CE │ +└────────────────────────┬─────────┘ └──────────────────────────────┘ + │ + ▼ +┌──────────────────────────────────────────────────────────────────────┐ +│ C51 categorical output: probs[B, 9, N_atoms] │ +│ • Adaptive atom support [v_min, v_max] from Pearl-1 controller │ +│ • ISV-driven instead of hardcoded │ +└────────────────────────────────┬─────────────────────────────────────┘ + │ + ▼ +┌──────────────────────────────────────────────────────────────────────┐ +│ GPU Thompson selector (existing) + alpha_confidence threshold gate │ +│ • Inverse-CDF over each expert's mixed probs │ +│ • Mapped-pinned action output (current implementation, unchanged) │ +└──────────────────────────────────────────────────────────────────────┘ +``` + +### Components (justification) + +| Component | Production kernel | Phase E role | Compat note | +|---|---|---|---| +| Sliding-window state buffer | new (CPU or GPU circular buffer) | Provides K bars of context per step | New buffer mgmt; reset on episode boundary | +| VSN | `vsn_feature_selection_kernel.cu` | Per-state feature-importance gating | Use 10 features instead of 6 groups; mask EMA into ISV | +| Mamba2 SSM | `mamba2_temporal_kernel.cu` | Temporal encoder | Same kernel; D=10 input, configure hidden width | +| GRN | `grn_kernel.cu` | Gated trunk encoding | Direct reuse | +| MoE gate | `moe_kernels.cu` + `moe_lambda_eff_kernel.cu` | Regime-aware Q-head dispatch | K_e=4 (matches spread-quintile regime structure) | +| C51 head | `alpha_c51.cu` (existing) | Per-expert categorical output | Already works | +| Aux trunk | `aux_trunk_forward/backward_kernel.cu` | Separate trunk for aux supervision | Stop-grad at encoder per `pearl_separate_aux_trunk_when_shared_starves` | +| Aux heads | `aux_heads_kernel.cu` + `aux_heads_loss_ema_kernel.cu` | Next-bar MSE + regime CE | Adapt label-builders to Phase E targets | +| Pearl-1 atom span | `pearl_1_atom_kernel.cu` | Adaptive [v_min, v_max] for C51 | Replaces hardcoded support; ISV-driven | +| Thompson selector | `alpha_c51.cu` (existing) | Action selection | Reads gate-mixed probs; no kernel change | + +### TFT correspondence + +| TFT component | Provided by | +|---|---| +| Variable Selection Network | `vsn_feature_selection_kernel.cu` (direct) | +| Gated Residual Network | `grn_kernel.cu` (direct) | +| Static covariate encoder | absent — Phase E has no static covariates; OK | +| LSTM seq-to-seq encoder | **substituted by Mamba2 SSM** (better; subquadratic) | +| Interpretable multi-head attention | optional Phase E.4.E follow-up via `attention_kernel.cu` | +| Quantile output | **substituted by C51 categorical** (already working) | + +The proposed architecture is structurally a TFT, with two strict upgrades vs the original Lim et al. 2021 design: +1. Mamba2 instead of LSTM (linear-time state accumulation, no vanishing-gradient). +2. C51 categorical output instead of quantile loss (proven calibration win). + +## 4. State representation changes + +Current Phase E state vector (10-dim, single timestep): +``` +[alpha_logit, alpha_confidence, spread_bps, l1_imbalance, ofi_sum_5, + mid_drift_5, position, step_normalized, log_tau, log_event_rate] +``` + +Proposed: same 10 features but **K-bar history** for all features. Window length K starts at 16 (~16 bars of context); validate empirically against K=32, K=64. Mamba2 handles long-horizon well (Phase 1d.2 used K=6000); the cost is GPU memory not architectural limit. + +Per-step input shape: `[B, K, 10]`. The buffer is maintained as a circular array; reset on episode start zeros it (or seeds with the first K observed states once available). + +## 5. Training changes + +- **Forward**: per step, push current state to buffer, run window → VSN → Mamba2 → GRN → (MoE → C51 head) + (aux trunk → aux heads). +- **Backward**: Q-loss (C51 CE) flows through MoE→GRN→Mamba2→VSN. Aux-loss flows through aux trunk only (stop-grad at encoder boundary, per `pearl_separate_aux_trunk_when_shared_starves`). +- **Loss balance**: per `pearl_loss_balance_controller`, signal-modulated target × Wiener-α. Phase E currently uses Q-loss only; aux-loss weight starts at 1.0 and is controller-driven. +- **Optimizer**: keep plain SGD initially (C51 works with SGD). Switch to Adam in Phase E.4.D if signal plateaus. +- **PER**: NOT in scope for E.4 (current on-policy training works); revisit only if learning is sample-starved. + +## 6. Phasing + +Each phase produces a runnable smoke + backtest. Each phase ships as a discrete experiment we can A/B against the prior best. + +**Phase E.4.A — Foundation (1 week)** +- Phase E.4.A.1: Sliding-window state buffer (GPU-resident, circular) +- Phase E.4.A.2: Hybrid MBP-10 depth loader — extend + `load_snapshots_from_fxcache` to peek MBP-10 by timestamp and + populate real L4-L10 bid/ask in `SnapshotRow.bid_l[3..10]` / + `ask_l[3..10]`. Preserves alpha_cache alignment. +- Phase E.4.A.3: Extend state vector to include depth features + (L1-L10 cumulative size, spread-curve slope, etc.) +- Phase E.4.A.4: Mamba2 forward + backward (already exists; wire into Phase E) +- Phase E.4.A.5: GRN trunk (already exists; wire) +- Phase E.4.A.6: C51 head reads h_s2 instead of state directly +- Phase E.4.A.7: **ISV-continual-learning toggle** — controllers fire + at eval time too (already in smoke at training; lift to backtest eval). +- Smoke + backtest vs C51-flat-baseline (target: maintain +10.4 at cost=0, lift half-tick) +- **Falsification: if Sharpe at cost=0 drops, the temporal architecture isn't lifting — pause and diagnose before adding more.** + +**Phase E.4.B — Regime gating (1 week)** +- Add VSN feature selection +- Add MoE 4-expert gating + λ controller (existing kernel) +- Validate expert utilization (should specialize across spread quintiles) +- A/B vs E.4.A + +**Phase E.4.C — Aux supervision (3-4 days)** +- Add aux trunk + heads (next-bar return MSE, 5-class regime CE) +- Validate aux losses converge (`pearl_separate_aux_trunk_when_shared_starves` gate: CE < 0.1 AND dir_acc > 0.95) +- Validate Q-learning isn't destabilized +- A/B vs E.4.B + +**Phase E.4.D — Adaptive atoms + polish (3-4 days)** +- Pearl-1 atom span controller (replaces hardcoded [-10, +10]) +- Add health composition (ISV[12]) feeding controller / atom span +- Trade-rate target sweep at high-cost regime (closes the trade-count economics) +- Final 2D sweep, write close-out memo + +**Phase E.4.E — Optional follow-ups (defer)** +- TFT interpretable multi-head attention +- IQN quantile head composed with C51 +- Pearl-4 Adam adaptive hyperparams + +## 7. Falsification criteria + +For the architecture upgrade to be worth the engineering cost: + +**Smoke must show (vs C51-flat-baseline R_mean = -1.1):** +- R_mean improvement ≥ 50% (i.e., R_mean ≥ -0.5) +- rvr maintained (≥ +1.04σ) +- Action entropy NOT collapsed to Wait-only (>0.5 × ln(9)) + +**Backtest must show (vs C51-flat-baseline Sharpe_ann at half-tick = -13.8):** +- Half-tick Sharpe_ann ≥ -8 (closes 5pt+ of the remaining 10pt gap to Phase 1d.4 baseline at -4.0) +- Trade rate ≤ 70/ep at best τ (moving toward Phase 1d.4's 20-50/ep regime) +- Cost=0 Sharpe_ann ≥ +8 (not regressed below the C51-flat result) + +If these don't hit by Phase E.4.B end: rollback to C51-flat and reconsider. The temporal architecture is a STRATEGIC bet — if it doesn't deliver, we have a known-good fallback (the C51-flat policy already shipped in this session). + +## 8. Risks + +- **State-buffer memory**: K=64 × 10 features × float = 2.5KB per env-instance. For batch backtest with 500 episodes the working set is fine. K=600 (full smoke horizon) is 24KB/instance — also fine. +- **Mamba2 init**: Phase 1d.2 used random init + trainable end-to-end (`pearl_tlob_no_pretraining`). Should work for Phase E too. +- **Aux head label leak**: `pearl_trend_scanning_purged_cv_doesnt_sterilize_forward_features` warns that forward-window labels leak into forward-horizon targets. Mitigation: aux supervision uses NEXT BAR (single-step) target, not multi-bar trend-scanning label. +- **MoE expert collapse**: if all 4 experts learn the same C51 distribution, MoE adds parameters without benefit. Mitigation: `moe_lambda_eff_kernel.cu` λ controller penalizes gate entropy collapse; existing kernel does this. +- **Loss balance instability**: with Q-loss + 2 aux losses, the encoder receives 3 gradient streams. `pearl_loss_balance_controller` is the production answer; adopt at Phase E.4.C. + +## 9. Open questions + +- **Window length K**: start at 16, sweep 16/32/64. Mamba2 doesn't impose a ceiling. +- **MoE expert count K_e**: start at 4 (matches spread-quintile structure from Phase 1c). Validate by checking expert utilization per regime. +- **Aux head regime classes**: 5 classes for spread quintiles, OR 5 classes for volatility quintiles? Pick spread (alpha-conditional dimension from Phase 1c). +- **Should we keep the threshold gate?** The current `--train-threshold` is a CPU-side gate. Once the Q-net is temporal + regime-aware, the gate may be redundant. Test with `--train-threshold 0.0` in E.4.B. + +## 10. Out of scope / future research + +**Out of scope (Phase E.4 core):** +- Different markets (validated as not the bottleneck for ES futures spread economics) +- Action-space changes (pruning falsified; full 9 stays) +- PER / off-policy replay (current on-policy training has no observable sample-starvation) +- LoRA dynamic weights at inference (aggressive ISV-continual-learning variant — Phase E.5 if needed) + +**Future research (Phase E.5+ or production-tier):** +- **TGGN graph reasoning over L1-L10 depth** + (`ml-supervised/src/tgnn/`): foxhunt has an existing CPU-based + Temporal Graph Gated Network for HFT (Performance targets <500ns + graph build, <1μs GNN inference). The trainable adapter is + GPU-native (`crates/ml/src/tgnn/trainable_adapter.rs` using cuBLAS + GpuLinear + GpuAdamW) but the message-passing layer is CPU. + Honest assessment for Phase E.4: marginal benefit over the + TFT-Mamba2 stack — the 81-dim Block-S features already encode + LOB structure and Mamba2 captures temporal context. TGGN's + strongest value is in (a) multi-instrument cross-asset graphs, + (b) full MBP-10 depth utilization, (c) production HFT + sub-1μs latency. With Pillar A (L4-L10 input wired), the depth + case becomes stronger — revisit as Phase E.5 candidate. +- **TFT interpretable multi-head attention** — + `attention_kernel.cu` + `attention_backward_kernel.cu`. + Adds variable-importance via attention weights. Already deferred + in section 6 as Phase E.4.E. +- **IQN dual head composed with C51** — joint sampling + (`pearl_thompson_for_distributional_action_selection`). Trade-off: + +2-3× training overhead for marginal exploration gain. +- **Pearl-4 Adam adaptive hyperparams** — replace plain SGD. +- **xLSTM / KAN / Liquid alternative encoders** — adapters exist + in `crates/ml/src/hyperopt/adapters/`. Test only if Mamba2 + plateaus. +- **Multi-instrument extension** (ES + NQ + RTY graph for + cross-asset signals) — opens up TGGN's strongest use case. + +## 11. Approval gate + +**This is a design proposal, not yet an implementation plan.** Before invoking `superpowers:writing-plans` to produce a step-by-step implementation: + +- [x] User approved the TFT-Mamba2 architecture ("the idea is great + we use that", 2026-05-15) +- [x] TGGN clarified — foxhunt-specific Temporal Graph Gated + Networks (`ml-supervised/src/tgnn/`). Existing CPU + implementation + GPU adapter. Deferred to Phase E.5+ research + per Section 10. +- [x] L4-L10 depth source verified — DBN MBP-10 files contain real + L1-L10. Pillar A (Section 2.5) integrates them via hybrid + MBP-10 peek alongside fxcache. +- [x] ISV-continual-learning concept adopted as Pillar B + (Section 2.5) — core architectural feature, not optional. +- [ ] User confirms phasing (4 phases × ~3 weeks total) +- [ ] User confirms scope: stop after E.4.D? Or extend with E.4.E? +- [ ] User approves the falsification criteria (smoke + backtest thresholds) + +After approval, the writing-plans skill produces task-by-task TDD implementation steps for Phase E.4.A first.