foxhunt/crates/ml/src/cuda_pipeline/stacker_threshold_controller.cu

// crates/ml/src/cuda_pipeline/stacker_threshold_controller.cu
//
// Phase E.2 Task 16 (2026-05-15) — Stacker-threshold controller +
// Kelly-attenuation controller. Engagement-rate self-correction per
// `pearl_engagement_rate_self_correction`. Single block, single thread;
// runs once per rollout-end.
//
// ISV slots driven (see `alpha_isv_slots.rs`):
//
//   543  STACKER_THRESHOLD_INDEX        — controller output [0, 0.5]
//   545  TRADE_RATE_OBSERVED_EMA_INDEX  — diagnostic (Pearl A+D EMA)
//   546  STACKER_KELLY_ATTENUATION_INX  — controller output [0.1, 1.0]
//
// ISV slots read:
//
//   544  TRADE_RATE_TARGET_INDEX        — anchor (host-init, never reset)
//
// **Control law:**
//
//   Trade rate (P-controller on the rate error):
//     observed = trade_count / max(decisions, 1)
//     ISV[545]_new ← Pearl A + Pearl D Wiener-α with floor=0.4 on observed
//                    (the controller co-adapts with the policy, so the floor
//                    keeps the EMA responsive per
//                    `pearl_wiener_alpha_floor_for_nonstationary`)
//     err_rate = ISV[545]_new − ISV[544]                     (signed)
//     ISV[543] ← clamp(ISV[543] + k_threshold · err_rate, 0, 0.5)
//
//   Kelly attenuation (P-controller on Sharpe error, floored):
//     err_sharpe = rollout_sharpe − target_sharpe
//     ISV[546] ← clamp(ISV[546] + k_atten · err_sharpe, 0.1, 1.0)
//
//   `k_threshold = 0.01`, `k_atten = 0.005` per plan defaults. Floored at
//   the lower clamp bound per `pearl_blend_formulas_must_have_permanent_floor`
//   (the Kelly attenuation can't be 0 — that would zero out position sizing
//   permanently; floor at 0.1 keeps the policy minimally active).
//
// **Wiener-α inline** (Pearl A + Pearl D with floor): same math as
// `apply_pearls_ad_kernel`, with `alpha_meta` overridden to a higher
// constant for the variance updates because the rate observation has
// O(0.1) range vs SP4's O(1) — needs faster variance tracking.

#include <cuda_runtime.h>

extern "C" __global__ void stacker_threshold_controller_update(
    /* Per-rollout raw inputs (mapped-pinned host scalars). */
    float        rollout_trade_count,
    float        rollout_total_decisions,
    float        rollout_realized_sharpe,
    /* Kernel-arg controller parameters. */
    float        target_sharpe,
    float        k_threshold,         // typ. 0.01
    float        k_atten,              // typ. 0.005
    float        wiener_alpha_floor,   // 0.4 per pearl_wiener_alpha_floor_for_nonstationary
    float        alpha_meta,           // meta-EMA rate for Wiener variance updates
    /* ISV slot indices passed as args (decouple slot numbering from kernel). */
    int          stacker_threshold_idx,
    int          trade_rate_target_idx,
    int          trade_rate_observed_idx,
    int          stacker_kelly_atten_idx,
    /* T16 vol-regime slots. Set both to a negative number to disable
       the regime scale (backward-compatible). */
    int          regime_vol_ema_idx,
    int          regime_vol_ref_idx,
    float        regime_scale_floor,   // typ. 0.25 — bottoms the multiplicative defense
    /* T16 vol_ref floor controller (set indices < 0 to disable). At
       cell-end, observe the filtered cell-min vol_obs and slow-EMA
       the floor anchor toward `floor_target_ratio × cell_min`. */
    int          regime_vol_ref_floor_idx,
    int          regime_vol_obs_min_idx,
    float        floor_target_ratio,   // typ. 0.5 — floor sits half the cell-min
    float        floor_update_rate,    // typ. 0.1 — slow EMA toward target
    float        vol_obs_min_sentinel, // typ. 1e10 — large value reset by host
    /* Device buffers. */
    float*       __restrict__ isv,
    float*       __restrict__ wiener_state /* [sample_var, diff_var, x_lag] for slot 545 */
) {
    if (blockIdx.x != 0 || threadIdx.x != 0) return;

    constexpr float EPS_DIV = 1.0e-8f;
    constexpr float THRESH_LO = 0.0f;
    constexpr float THRESH_HI = 0.5f;
    constexpr float ATTEN_LO = 0.1f;
    constexpr float ATTEN_HI = 1.0f;

    // ---- (1) Observed trade rate ----
    const float observed_rate =
        rollout_trade_count / fmaxf(rollout_total_decisions, 1.0f);

    // ---- (2) Pearl A + Pearl D update on ISV[545] (observed-rate EMA) ----
    //
    // Pearl A: sentinel-bootstrap. If both prev_x_mean == 0 and x_lag == 0
    // (untouched slot or fold-boundary reset), replace directly with the
    // observation. Wiener state initialised to (sample_var=0, diff_var=0,
    // x_lag=observation).
    //
    // Pearl D: variance-EMA at alpha_meta; α* = diff_var / (diff_var +
    // sample_var + EPS_DIV); EMA blend `(1-α)·prev + α·obs`. Then APPLY
    // FLOOR: `α* = max(α*, wiener_alpha_floor)` per
    // `pearl_wiener_alpha_floor_for_nonstationary` — the controller
    // co-adapts with the policy so we can't let α* drift to ~0 on a
    // briefly-stationary segment.
    const float prev_x_mean = isv[trade_rate_observed_idx];
    const float sample_var = wiener_state[0];
    const float diff_var = wiener_state[1];
    const float x_lag = wiener_state[2];

    float new_x_mean;
    float new_sample_var;
    float new_diff_var;
    float new_x_lag;
    if (prev_x_mean == 0.0f && x_lag == 0.0f) {
        // Pearl A bootstrap.
        new_x_mean = observed_rate;
        new_sample_var = 0.0f;
        new_diff_var = 0.0f;
        new_x_lag = observed_rate;
    } else {
        const float sample_diff = observed_rate - prev_x_mean;
        const float x_lag_diff = observed_rate - x_lag;
        new_sample_var = (1.0f - alpha_meta) * sample_var
                       + alpha_meta * sample_diff * sample_diff;
        new_diff_var = (1.0f - alpha_meta) * diff_var
                     + alpha_meta * x_lag_diff * x_lag_diff;
        const float alpha_opt_raw =
            new_diff_var / (new_diff_var + new_sample_var + EPS_DIV);
        const float alpha_opt = fmaxf(alpha_opt_raw, wiener_alpha_floor);
        new_x_mean = (1.0f - alpha_opt) * prev_x_mean + alpha_opt * observed_rate;
        new_x_lag = observed_rate;
    }
    isv[trade_rate_observed_idx] = new_x_mean;
    wiener_state[0] = new_sample_var;
    wiener_state[1] = new_diff_var;
    wiener_state[2] = new_x_lag;

    // ---- (3) Update STACKER_THRESHOLD (slot 543) from rate error ----
    const float target_rate = isv[trade_rate_target_idx];
    const float err_rate = new_x_mean - target_rate;
    const float old_threshold = isv[stacker_threshold_idx];
    const float new_threshold =
        fmaxf(THRESH_LO, fminf(THRESH_HI, old_threshold + k_threshold * err_rate));
    isv[stacker_threshold_idx] = new_threshold;

    // ---- (4) Update KELLY_ATTENUATION (slot 546) from Sharpe error ----
    //
    // Pearl A bootstrap: if slot is 0 (sentinel — untouched / fold-reset),
    // start at the ceiling (1.0) so the policy is fully aggressive from the
    // first rollout and the controller dials it down from there. NOT
    // floored at sentinel — sentinel is the START state, the floor is the
    // operational floor.
    const float prev_atten_raw = isv[stacker_kelly_atten_idx];
    const float prev_atten = prev_atten_raw == 0.0f ? ATTEN_HI : prev_atten_raw;
    const float err_sharpe = rollout_realized_sharpe - target_sharpe;
    const float reactive_atten =
        fmaxf(ATTEN_LO, fminf(ATTEN_HI, prev_atten + k_atten * err_sharpe));

    // ---- (4b) T16 pre-emptive vol-regime scale ----
    //
    // The reactive Sharpe-error controller above only fires AFTER losses
    // materialize. The vol-regime scale fires AS SOON AS realized vol
    // diverges from the trained-on reference vol. The two compose
    // multiplicatively — reactive sets the upper bound based on observed
    // P&L, regime trims further when conditions look unfriendly.
    //
    // regime_scale = clamp(vol_ref / max(vol_ema, ε), regime_scale_floor, 1.0)
    //
    // vol_ema << vol_ref (calm market vs trained turbulence) → scale = 1.0 (no trim)
    // vol_ema ≈ vol_ref (same regime as training)              → scale ≈ 1.0
    // vol_ema = 2·vol_ref (vol spike)                          → scale = 0.5
    // vol_ema = 4·vol_ref (extreme spike)                      → scale = 0.25 (floor)
    //
    // Disabled when either slot index is negative OR when vol_ref ≤ 0
    // (regime kernel hasn't bootstrapped yet — must not punish cold start).
    float regime_scale = 1.0f;
    if (regime_vol_ema_idx >= 0 && regime_vol_ref_idx >= 0) {
        const float vol_ema = isv[regime_vol_ema_idx];
        const float vol_ref = isv[regime_vol_ref_idx];
        if (vol_ref > 0.0f && vol_ema > 0.0f) {
            const float raw_scale = vol_ref / vol_ema;
            regime_scale = fmaxf(regime_scale_floor, fminf(1.0f, raw_scale));
        }
    }

    const float final_atten = fmaxf(ATTEN_LO, reactive_atten * regime_scale);
    isv[stacker_kelly_atten_idx] = final_atten;

    // ---- (5) T16 vol_ref floor controller update ----
    //
    // The vol_ref floor anchor (slot 552) prevents the deadband
    // deadlock; it has to be SIZED to the instrument's typical
    // vol_obs range. Initial seed is a host hyperparam; this
    // controller refines it from observed signal so the same binary
    // works across instruments / time regimes without re-tuning.
    //
    // Signal: the trained-regime baseline vol_ref itself (slot 550).
    // Target: `floor_target_ratio × vol_ref` — floor sits at a small
    // fraction of the regime baseline (typically 0.1 = 10%). This is
    // architecturally coherent: floor is a fraction of the same signal
    // the spike detector compares against (vol_ema vs vol_ref), so the
    // floor scales naturally with regime shifts as vol_ref tracks the
    // current regime.
    // Update: slow EMA `floor_new = floor_old + α × (target − old)`.
    //
    // The per-step min(vol_obs) slot 553 is now an observability-only
    // diagnostic — written by the regime kernel for inspection but no
    // longer consumed by the controller (the 0.5 × cell_min target
    // overshot in walk-forward and effectively disabled the defense).
    //
    // Disabled when either floor or vol_ref index < 0
    // (backward-compat for smoke + kernel unit test).
    (void)regime_vol_obs_min_idx;  // unused signal channel — see above
    (void)vol_obs_min_sentinel;
    if (regime_vol_ref_floor_idx >= 0 && regime_vol_ref_idx >= 0) {
        const float vol_ref_current = isv[regime_vol_ref_idx];
        if (vol_ref_current > 1.0e-12f) {
            const float floor_prev = isv[regime_vol_ref_floor_idx];
            const float target = floor_target_ratio * vol_ref_current;
            const float floor_new =
                floor_prev + floor_update_rate * (target - floor_prev);
            isv[regime_vol_ref_floor_idx] = fmaxf(floor_new, 1.0e-12f);
        }
    }

    __threadfence_system();
}