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

feat(cuda): add dsr_step() device function for GPU DSR reward shaping

Browse Source 
Co-Authored-By: Claude Sonnet 4.6 <noreply@anthropic.com>
This commit is contained in:
jgrusewski
2026-03-10 12:44:01 +01:00
parent 9a96691001
commit 61a9f45af5
1 changed files with 237 additions and 0 deletions
Download Patch File Download Diff File Expand all files Collapse all files

237
crates/ml/src/cuda_pipeline/common_device_functions.cuh
View File

@@ -777,3 +777,240 @@ __device__ float curiosity_inference(
return fminf(mse, max_reward);
}
/* ------------------------------------------------------------------ */
/* Differential Sharpe Ratio (Moody & Saffell 2001) */
/* ------------------------------------------------------------------ */
/**
* Compute one DSR step and update EMA accumulators in-place.
*
* D_t = (B_prev * delta_A - 0.5 * A_prev * delta_B) / (B_prev - A_prev^2)^{3/2}
*
* where delta_A = r_t - A_prev, delta_B = r_t^2 - B_prev.
* Returns clamped DSR in [-5, 5].
*/
__device__ __forceinline__ float dsr_step(
float r_t,
float* ema_A, /* EMA of returns (in/out) */
float* ema_B, /* EMA of squared returns (in/out) */
int* dsr_init, /* 0 = first call, set to 1 after init */
float eta /* EMA decay rate (e.g. 0.01) */
) {
if (!(*dsr_init)) {
*ema_A = r_t;
*ema_B = r_t * r_t + 1e-8f; /* avoid zero variance */
*dsr_init = 1;
return 0.0f; /* No DSR on first step */
}
float a_prev = *ema_A;
float b_prev = *ema_B;
/* Update EMAs */
*ema_A = a_prev + eta * (r_t - a_prev);
*ema_B = b_prev + eta * (r_t * r_t - b_prev);
/* DSR formula */
float delta_a = r_t - a_prev;
float delta_b = r_t * r_t - b_prev;
float variance = b_prev - a_prev * a_prev;
if (variance < 1e-12f) {
/* Fallback: scale raw PnL to approximate DSR magnitude range */
return fmaxf(-5.0f, fminf(5.0f, r_t * 100.0f));
}
float denom = powf(variance, 1.5f);
float dsr = (b_prev * delta_a - 0.5f * a_prev * delta_b) / denom;
return fmaxf(-5.0f, fminf(5.0f, dsr));
}
/* ------------------------------------------------------------------ */
/* Order Routing (GPU-side OrderRouter::route) */
/* ------------------------------------------------------------------ */
/**
* Deterministic order routing based on market microstructure.
*
* Maps (exposure_idx, spread, volatility) → (order_type, urgency).
* Matches CPU `OrderRouter::route()` in ml-core/src/order_router.rs exactly.
*
* Order types: 0=LimitMaker, 1=IoC, 2=Market
* Urgency: 0=Patient, 1=Normal, 2=Aggressive
*/
__device__ __forceinline__ void route_order(
float spread,
float median_spread,
float volatility,
float median_vol,
int* out_order_type,
int* out_urgency
) {
/* Order type selection based on spread vs median */
int order;
if (median_spread <= 0.0f) {
order = 2; /* Market — no spread reference */
} else if (spread < median_spread) {
order = 0; /* LimitMaker — tight spread, save costs */
} else if (spread > 2.0f * median_spread) {
order = 2; /* Market — wide spread, ensure fill */
} else {
order = 1; /* IoC — moderate spread */
}
/* Urgency selection based on volatility vs median */
int urgency;
if (median_vol <= 0.0f) {
urgency = 1; /* Normal — no vol reference */
} else if (volatility > 1.5f * median_vol) {
urgency = 2; /* Aggressive — high vol */
} else if (volatility < 0.5f * median_vol) {
urgency = 0; /* Patient — low vol */
} else {
urgency = 1; /* Normal */
}
*out_order_type = order;
*out_urgency = urgency;
}
/**
* Order-type-dependent transaction cost rate.
*
* Matches CPU FillSimulator::spread_cost() cost structure:
* Market = spread_cost_frac * spread_bps / 10000
* IoC = spread_cost_frac * 0.5 * spread_bps / 10000
* LimitMaker = -(spread_capture_frac * spread_bps / 10000) [rebate]
*
* Returns cost as a fraction of trade value (positive = cost, negative = rebate).
*/
__device__ __forceinline__ float order_type_tx_cost(
int order_type,
float spread_bps,
float spread_cost_frac,
float spread_capture_frac
) {
switch (order_type) {
case 0: /* LimitMaker — maker rebate */
return -(spread_bps * spread_capture_frac / 10000.0f);
case 1: /* IoC — half spread cost */
return spread_bps * spread_cost_frac * 0.5f / 10000.0f;
case 2: /* Market — full spread cost */
return spread_bps * spread_cost_frac / 10000.0f;
default:
return spread_bps * spread_cost_frac / 10000.0f;
}
}
/* ------------------------------------------------------------------ */
/* Fill Simulation (GPU-side FillSimulator::simulate_fill) */
/* ------------------------------------------------------------------ */
/**
* Splitmix64-style deterministic hash: (step, action_idx) → [0.0, 1.0).
*
* Bit-exact match with CPU splitmix64_unit() in fill_simulator.rs.
* Uses the same Weyl-sequence seed combining and two-round finalizer.
*/
__device__ __forceinline__ float splitmix64_fill(int step, int action_idx) {
/* Combine step and action_idx with Weyl sequence constants */
unsigned long long z = (unsigned long long)step * 0x9e3779b97f4a7c15ULL
+ (unsigned long long)action_idx * 0x6a09e667f3bcc908ULL;
/* Splitmix64 two-round finalizer */
z = (z ^ (z >> 30)) * 0xbf58476d1ce4e5b9ULL;
z = (z ^ (z >> 27)) * 0x94d049bb133111ebULL;
z ^= (z >> 31);
/* Convert to [0.0, 1.0) — upper 53 bits as double mantissa, then narrow to float */
double d = (double)(z >> 11) / (double)(1ULL << 53);
return (float)d;
}
/**
* Compute fill probability for a given order type and market conditions.
*
* Matches CPU FillSimulator::fill_probability() exactly.
*
* Parameters:
* order_type: 0=LimitMaker, 1=IoC, 2=Market
* urgency: 0=Patient, 1=Normal, 2=Aggressive
* normalized_vol: volatility / median_volatility (1.0 = average)
* ioc_fill_prob: base IoC fill probability (default 0.85)
* limit_fill_min: minimum limit fill probability (default 0.30)
* limit_fill_max: maximum limit fill probability (default 0.80)
*/
__device__ __forceinline__ float fill_probability(
int order_type,
int urgency,
float normalized_vol,
float ioc_fill_prob,
float limit_fill_min,
float limit_fill_max
) {
float base_prob;
switch (order_type) {
case 2: /* Market — always fills */
base_prob = 1.0f;
break;
case 1: /* IoC */
base_prob = ioc_fill_prob;
break;
case 0: /* LimitMaker — vol-dependent */
default: {
float vol_clamped = fminf(fmaxf(normalized_vol, 0.0f), 3.0f);
float vol_factor = vol_clamped / 3.0f;
base_prob = limit_fill_min + vol_factor * (limit_fill_max - limit_fill_min);
break;
}
}
/* Urgency modifier */
float urgency_mod;
switch (urgency) {
case 0: urgency_mod = 0.85f; break; /* Patient */
case 2: urgency_mod = 1.10f; break; /* Aggressive */
default: urgency_mod = 1.0f; break; /* Normal */
}
float prob = base_prob * urgency_mod;
return fminf(fmaxf(prob, 0.0f), 1.0f);
}
/**
* Simulate order fill using deterministic hash.
*
* Matches CPU FillSimulator::simulate_fill() exactly.
*
* Returns: 1 if filled, 0 if not filled.
* Writes cost adjustment to *out_cost_adj (positive=cost, negative=rebate, 0=no fill).
*/
__device__ __forceinline__ int simulate_fill_check(
int order_type,
int urgency,
float normalized_vol,
float spread_bps,
int step,
int action_idx,
float ioc_fill_prob,
float limit_fill_min,
float limit_fill_max,
float spread_cost_frac,
float spread_capture_frac,
float* out_cost_adj
) {
float prob = fill_probability(order_type, urgency, normalized_vol,
ioc_fill_prob, limit_fill_min, limit_fill_max);
float pseudo_random = splitmix64_fill(step, action_idx);
int filled = (pseudo_random < prob) ? 1 : 0;
if (filled) {
*out_cost_adj = order_type_tx_cost(order_type, spread_bps,
spread_cost_frac, spread_capture_frac);
} else {
*out_cost_adj = 0.0f;
}
return filled;
}