From c74a687ea8ca316da2fcd700ca0fcf1090abb689 Mon Sep 17 00:00:00 2001 From: jgrusewski Date: Thu, 16 Apr 2026 21:56:44 +0200 Subject: [PATCH] =?UTF-8?q?feat:=20position-gated=20episodes=20+=205000-ba?= =?UTF-8?q?r=20limit=20=E2=80=94=20close=2045x=20training/val=20gap?= MIME-Version: 1.0 Content-Type: text/plain; charset=UTF-8 Content-Transfer-Encoding: 8bit Episode done flag: timer-based -> position-gated (trade complete = done). V(flat)=0 is correct terminal anchor. Soft reset keeps equity on trade completion; hard reset only on data-end or capital breach. H100: 100 bars -> 5000 bars, gpu_n_episodes -> 1024. ExperienceProfile gains optional gpu_n_episodes field. Co-Authored-By: Claude Opus 4.6 (1M context) --- config/gpu/h100.toml | 3 +- crates/ml-core/src/gpu/profile.rs | 7 +- crates/ml/examples/train_baseline_rl.rs | 1 + .../src/cuda_pipeline/experience_kernels.cu | 66 ++++++++++++------- 4 files changed, 53 insertions(+), 24 deletions(-) diff --git a/config/gpu/h100.toml b/config/gpu/h100.toml index 0d6246e85..60b2210dd 100644 --- a/config/gpu/h100.toml +++ b/config/gpu/h100.toml @@ -7,7 +7,8 @@ hidden_dim_base = 256 replay_buffer_vram_fraction = 0.70 [experience] -gpu_timesteps_per_episode = 100 +gpu_timesteps_per_episode = 5000 +gpu_n_episodes = 1024 [cuda] cuda_stack_bytes = 65536 # 64KB diff --git a/crates/ml-core/src/gpu/profile.rs b/crates/ml-core/src/gpu/profile.rs index b94e020f4..89b0897f8 100644 --- a/crates/ml-core/src/gpu/profile.rs +++ b/crates/ml-core/src/gpu/profile.rs @@ -47,6 +47,9 @@ pub struct TrainingProfile { #[derive(Debug, Clone, Deserialize)] pub struct ExperienceProfile { pub gpu_timesteps_per_episode: usize, + /// Number of parallel episodes (optional — training profile overrides if set). + #[serde(default)] + pub gpu_n_episodes: Option, } /// CUDA kernel configuration. @@ -203,6 +206,7 @@ impl GpuProfile { }, experience: ExperienceProfile { gpu_timesteps_per_episode: 200, + gpu_n_episodes: None, }, cuda: CudaProfile { cuda_stack_bytes: 0x8000, @@ -284,7 +288,8 @@ mod tests { assert_eq!(profile.training.batch_size, 8192); assert_eq!(profile.training.num_atoms, 51); assert_eq!(profile.training.buffer_size, 500_000); - assert_eq!(profile.experience.gpu_timesteps_per_episode, 100); + assert_eq!(profile.experience.gpu_timesteps_per_episode, 5000); + assert_eq!(profile.experience.gpu_n_episodes, Some(1024)); assert_eq!(profile.cuda.cuda_stack_bytes, 65536); } diff --git a/crates/ml/examples/train_baseline_rl.rs b/crates/ml/examples/train_baseline_rl.rs index 46f3224bd..934818f73 100644 --- a/crates/ml/examples/train_baseline_rl.rs +++ b/crates/ml/examples/train_baseline_rl.rs @@ -424,6 +424,7 @@ fn build_dqn_hyperparams( dataset_path: args.dataset_path.as_ref().map(|p| p.to_string_lossy().into_owned()), replay_buffer_vram_fraction: gpu_profile.training.replay_buffer_vram_fraction, gpu_timesteps_per_episode: gpu_profile.experience.gpu_timesteps_per_episode, + gpu_n_episodes: gpu_profile.experience.gpu_n_episodes.unwrap_or(256), ..DQNHyperparameters::default() }; diff --git a/crates/ml/src/cuda_pipeline/experience_kernels.cu b/crates/ml/src/cuda_pipeline/experience_kernels.cu index 70c776976..a2b909117 100644 --- a/crates/ml/src/cuda_pipeline/experience_kernels.cu +++ b/crates/ml/src/cuda_pipeline/experience_kernels.cu @@ -1755,7 +1755,13 @@ extern "C" __global__ void experience_env_step( /* ---- Done detection ---- */ int next_bar = bar_idx + 1; - int done = (next_bar >= total_bars || check_capital_floor(new_portfolio_value, peak_equity)) ? 1 : 0; + /* Position-gated done: episode ends when model completes a trade cycle + * (was positioned -> now flat). V(flat)=0 is correct terminal anchor. + * Also done at end-of-data or capital floor breach (existing). */ + int trade_complete = (fabsf(pre_trade_position) > 0.001f && fabsf(position) < 0.001f) ? 1 : 0; + int capital_breach = check_capital_floor(new_portfolio_value, peak_equity); + int data_end = (next_bar >= total_bars) ? 1 : 0; + int done = trade_complete | capital_breach | data_end; /* #19 Position entropy bonus at episode end — all 4 branches */ if (done && position_histogram != NULL && position_entropy_weight > 0.0f) { @@ -1960,28 +1966,44 @@ extern "C" __global__ void experience_env_step( current_timesteps[i] = t + 1; /* ---- Episode reset on done ---- */ - /* When done=1, reset portfolio to fresh capital for the next episode. - * Without this, the blown portfolio persists and every subsequent step - * hits the pre-trade floor check, producing stuck episodes with fake MaxDD. */ + /* Position-gated episodes: three done reasons, two reset paths. + * - data_end / capital_breach: HARD reset (full restart with fresh capital) + * - trade_complete: SOFT reset (keep equity, clear trade-tracking only) + * Soft reset lets the model accumulate gains/losses across trade cycles + * within a single data window, preventing the 45x train/val gap caused by + * resetting equity after every trade completion. */ if (done) { - float init_cap = (peak_equity); /* use peak as starting capital for next ep */ - ps[0] = 0.0f; /* position = flat */ - ps[1] = init_cap; /* cash */ - ps[2] = init_cap; /* portfolio_value */ - ps[7] = init_cap; /* peak_equity */ - ps[8] = 0.0f; /* flat_counter */ - ps[9] = init_cap; /* prev_equity */ - ps[10] = 0.0f; /* hold_time */ - ps[11] = 0.0f; /* realized_pnl */ - ps[12] = 0.0f; /* entry_price */ - ps[13] = 0.0f; /* trade_start_pnl */ - ps[14] = 0.0f; /* win/loss/Kelly counters */ - ps[15] = 0.0f; - ps[16] = 0.0f; - ps[17] = 0.0f; - ps[18] = 0.0f; - ps[19] = 0.0f; - current_timesteps[i] = 0; + if (data_end || capital_breach) { + /* Hard reset: new data window or capital wipeout — full restart */ + float init_cap = (peak_equity); /* use peak as starting capital for next ep */ + ps[0] = 0.0f; /* position = flat */ + ps[1] = init_cap; /* cash */ + ps[2] = init_cap; /* portfolio_value */ + ps[7] = init_cap; /* peak_equity */ + ps[8] = 0.0f; /* flat_counter */ + ps[9] = init_cap; /* prev_equity */ + ps[10] = 0.0f; /* hold_time */ + ps[11] = 0.0f; /* realized_pnl */ + ps[12] = 0.0f; /* entry_price */ + ps[13] = 0.0f; /* trade_start_pnl */ + ps[14] = 0.0f; /* win/loss/Kelly counters */ + ps[15] = 0.0f; + ps[16] = 0.0f; + ps[17] = 0.0f; + ps[18] = 0.0f; + ps[19] = 0.0f; + current_timesteps[i] = 0; + } else { + /* Soft reset (trade_complete): keep equity, clear trade state only. + * The model starts next segment with accumulated gains/losses. + * Position is already 0 (model chose Flat). Cash, equity, + * win/loss counters, DSR state — all preserved. */ + ps[10] = 0.0f; /* hold_time — new trade starts fresh */ + ps[12] = 0.0f; /* entry_price — no active trade */ + ps[13] = (ps[11]); /* trade_start_pnl = current realized_pnl */ + ps[20] = 0.0f; /* intra_trade_max_dd — reset for next trade */ + current_timesteps[i] = 0; + } } }