keyhog-scanner 0.5.73

//! ML-based secret scoring with a tiny mixture-of-experts network.
//!
//! Architecture: gate Linear(D,6) → Softmax plus 6 experts of
//! Linear(D,32) → ReLU → Linear(32,16) → ReLU → Linear(16,1), then
//! a weighted logit sum followed by Sigmoid, where `D = NUM_FEATURES` (55).
//! Model weights are embedded in `ml_weights.rs` as little-endian f32 values.
//! Inference: typically under ~100μs per prediction on the test hardware
//!
//! The 55 input features capture value/context shape plus active-detector and
//! candidate-channel evidence: length,
//! entropy, char diversity, known prefixes, context keywords, placeholder
//! patterns, structural signals, coarse file-type cues, the decode-structure
//! verdict (feature #41, base64/hex → magic-bytes/protobuf) that drove the
//! base64-of-binary false-flag rate to 0%, and the keyword-specificity signal
//! (feature #42, context names a SPECIFIC service vs generic role words only
//! see `service_vocab`), which separates service-keyword+UUID real secrets
//! from generic-keyword+UUID identifiers. Features 43-54 describe the exact
//! TOML owner, producer channel, and detector-owned entropy family.

// Submodules live in `ml_scorer/` (native path resolution), matching the
// `foo.rs` + `foo/` layout used across the workspace (e.g. sources/filesystem).
// `model_arch` is the single owner of every layer dimension/param/offset const;
// `ml_weights` (buffer layout) and this module (forward pass) both import it.
pub(crate) mod ml_weights;
pub(crate) mod model_arch;

use std::cell::RefCell;

pub(crate) mod ml_features;
pub(crate) mod service_vocab;
#[cfg(test)]
pub(crate) mod service_vocab_build;
#[cfg(all(test, feature = "ml"))]
pub(crate) use ml_features::compute_features_public;
pub use ml_features::compute_features_with_config;
pub(crate) use ml_features::NUM_FEATURES;
pub use ml_features::{compute_features_for_detector_with_config, MlCandidateChannel};

/// Borrowed model input that lets batch inference consume its owning queue directly.
#[cfg(feature = "ml")]
pub(crate) trait MlScoreInput: Sync {
    fn ml_text(&self) -> &str;
    fn ml_features(&self, config: &crate::types::ScannerConfig) -> [f32; NUM_FEATURES];
}

#[cfg(feature = "ml")]
impl MlScoreInput for (&str, &str) {
    #[inline]
    fn ml_text(&self) -> &str {
        self.0
    }

    #[inline]
    fn ml_features(&self, config: &crate::types::ScannerConfig) -> [f32; NUM_FEATURES] {
        if self.0.is_empty() {
            return [0.0; NUM_FEATURES];
        }
        compute_features_with_config(
            self.0,
            self.1,
            &config.known_prefixes,
            &config.secret_keywords,
            &config.test_keywords,
            &config.placeholder_keywords,
        )
    }
}

#[cfg(feature = "ml")]
impl MlScoreInput for crate::types::MlPendingMatch {
    #[inline]
    fn ml_text(&self) -> &str {
        self.pending_raw_match.credential.as_ref()
    }

    #[inline]
    fn ml_features(&self, _config: &crate::types::ScannerConfig) -> [f32; NUM_FEATURES] {
        self.ml_features
    }
}

// Forward-pass dimensions + sigmoid clamp: imported from the single owner. The
// re-export keeps `crate::ml_scorer::SIGMOID_SATURATION` stable for `testing.rs`.
pub(crate) use model_arch::SIGMOID_SATURATION;
use model_arch::{EXPERT_COUNT, EXPERT_FC1_OUT, EXPERT_FC2_OUT};

/// Batch-size crossover for ML scoring. Below this, one fused serial loop
/// avoids Rayon split/join overhead. At or above it, feature extraction and
/// scoring use the configured CPU worker pool. This remains available without
/// the `ml` feature because the always-built quantized scorer shares the bound.
pub(crate) const ML_PARALLEL_BATCH_THRESHOLD: usize = 64;

/// Bounded per-thread score-cache capacity: the map is cleared wholesale once it
/// reaches this many entries (see `util_hash::memoize_by_hash`). 256 covers the
/// distinct matches of a single file's batch scoring. Single owner for the cap so
/// the number in the code and the "256-entry bounded cache" prose cannot drift.
pub(crate) const SCORE_CACHE_CAPACITY: usize = 256;

// NUM_FEATURES is `model_arch::INPUT_DIM` (one owner), so the forward pass's
// const-generic layer widths (EXPERT_FC1_OUT/EXPERT_FC2_OUT) and the buffer
// strides in `ml_weights` are the SAME consts, a retrain cannot make them
// disagree. `weights.bin`'s byte length is still checked at parse time.

/// Score a candidate secret and its surrounding context using default (empty) heuristic lists.
pub(crate) fn score(text: &str, context: &str) -> f64 {
    score_with_config(text, context, &[], &[], &[], &[])
}

/// Score a candidate secret and its surrounding context with provided heuristic lists.
pub(crate) fn score_with_config(
    text: &str,
    context: &str,
    known_prefixes: &[String],
    secret_keywords: &[String],
    test_keywords: &[String],
    placeholder_keywords: &[String],
) -> f64 {
    if text.is_empty() {
        return 0.0;
    }

    thread_local! {
        // FNV-1a keyed cache - ~100x faster than SHA-256 for cache lookups.
        // 256-entry bounded cache covers batch scoring of one file's matches.
        static SCORE_CACHE: RefCell<std::collections::HashMap<u64, f64>> =
            RefCell::new(std::collections::HashMap::with_capacity(64));
    }

    // The score depends on the four configured feature vocabularies as well as
    // text and context. Length-prefix every field so embedded NULs or list
    // boundaries cannot alias another input, and include every vocabulary so
    // scanners with different resolved configs never reuse each other's score.
    let cache_key = score_cache_key(
        text,
        context,
        known_prefixes,
        secret_keywords,
        test_keywords,
        placeholder_keywords,
    );

    // Shared bounded-cache idiom (`util_hash::memoize_by_hash`) instead of a
    // hand-rolled get / compute / clear-at-cap / insert. This is the migration
    // util_hash.rs's doc already names (it lists `ml_scorer::score_with_config`
    // as a memoize_by_hash consumer); the cap stays 256 (one file's matches),
    // and on a miss the same features→forward_pass→f64 is computed, so the
    // observable score is byte-identical to the hand-rolled cache.
    crate::util_hash::memoize_by_hash(&SCORE_CACHE, cache_key, SCORE_CACHE_CAPACITY, || {
        let features = compute_features_with_config(
            text,
            context,
            known_prefixes,
            secret_keywords,
            test_keywords,
            placeholder_keywords,
        );
        forward_pass(&features) as f64
    })
}

pub(crate) fn score_cache_key(
    text: &str,
    context: &str,
    known_prefixes: &[String],
    secret_keywords: &[String],
    test_keywords: &[String],
    placeholder_keywords: &[String],
) -> u64 {
    fn write_field(hasher: &mut crate::util_hash::FnvHasher, value: &[u8]) {
        hasher.write(&(value.len() as u64).to_le_bytes());
        hasher.write(value);
    }

    let mut hasher = crate::util_hash::FnvHasher::new();
    hasher.write(b"keyhog-ml-score-v2");
    write_field(&mut hasher, text.as_bytes());
    write_field(&mut hasher, context.as_bytes());
    for values in [
        known_prefixes,
        secret_keywords,
        test_keywords,
        placeholder_keywords,
    ] {
        hasher.write(&(values.len() as u64).to_le_bytes());
        for value in values {
            write_field(&mut hasher, value.as_bytes());
        }
    }
    hasher.finish()
}

/// Preserve pending-match cardinality before confidence blending. A malformed
/// backend score vector is a typed backend failure, never permission to drop
/// queued findings or silently rescore them.
#[cfg(feature = "ml")]
pub(crate) fn complete_batch_scores_with_config<T: MlScoreInput>(
    scores: Vec<f64>,
    inputs: &[T],
    _config: &crate::types::ScannerConfig,
) -> crate::Result<Vec<f64>> {
    if scores.len() != inputs.len() {
        return Err(crate::ScanError::Gpu(format!(
            "ML backend score count mismatch: expected {}, received {}",
            inputs.len(),
            scores.len()
        )));
    }
    Ok(scores)
}

#[cfg(feature = "ml")]
struct PreparedQuantizedBatch {
    scores: Vec<f64>,
    accelerated_candidate_ids: Vec<usize>,
    accelerated_rows: Vec<crate::confidence::quantized::QuantizedFeatureRow>,
    cpu_rows: Vec<(usize, Option<[f32; NUM_FEATURES]>)>,
}

#[cfg(feature = "ml")]
fn prepare_quantized_batch<T: MlScoreInput>(
    inputs: &[T],
    config: &crate::types::ScannerConfig,
) -> Result<PreparedQuantizedBatch, crate::confidence::quantized::QuantizedConfidenceError> {
    use crate::confidence::quantized::{
        candidate_score_ownership, CandidateScoreOwnership, QuantizedConfidenceError,
        QuantizedFeatureRow, MAX_CANDIDATES_PER_BATCH,
    };

    if inputs.len() > MAX_CANDIDATES_PER_BATCH {
        return Err(QuantizedConfidenceError::BatchTooLarge {
            candidates: inputs.len(),
            maximum: MAX_CANDIDATES_PER_BATCH,
        });
    }
    let mut accelerated_candidate_ids = Vec::new();
    let mut accelerated_rows = Vec::new();
    let mut cpu_rows = Vec::new();
    let mut scores = Vec::new();
    scores.try_reserve_exact(inputs.len()).map_err(|_| {
        QuantizedConfidenceError::BackendFailure(
            "quantized confidence score allocation failed".into(),
        )
    })?;
    scores.resize(inputs.len(), 0.0);
    accelerated_candidate_ids
        .try_reserve_exact(inputs.len())
        .map_err(|_| {
            QuantizedConfidenceError::BackendFailure(
                "quantized confidence candidate allocation failed".into(),
            )
        })?;
    accelerated_rows
        .try_reserve_exact(inputs.len())
        .map_err(|_| {
            QuantizedConfidenceError::BackendFailure(
                "quantized confidence feature allocation failed".into(),
            )
        })?;
    cpu_rows.try_reserve_exact(inputs.len()).map_err(|_| {
        QuantizedConfidenceError::BackendFailure(
            "CPU-owned confidence feature allocation failed".into(),
        )
    })?;
    for (candidate_id, input) in inputs.iter().enumerate() {
        let text = input.ml_text();
        match candidate_score_ownership(text.as_bytes()) {
            CandidateScoreOwnership::Cpu => cpu_rows.push((candidate_id, None)),
            CandidateScoreOwnership::Accelerated => {
                let features = input.ml_features(config);
                match QuantizedFeatureRow::from_float(&features) {
                    Ok(row) => {
                        accelerated_candidate_ids.push(candidate_id);
                        accelerated_rows.push(row);
                    }
                    Err(_) => cpu_rows.push((candidate_id, Some(features))),
                }
            }
        }
    }
    Ok(PreparedQuantizedBatch {
        scores,
        accelerated_candidate_ids,
        accelerated_rows,
        cpu_rows,
    })
}

#[cfg(feature = "ml")]
fn score_cpu_owned_rows<T: MlScoreInput>(
    batch: &mut PreparedQuantizedBatch,
    inputs: &[T],
    config: &crate::types::ScannerConfig,
) {
    for (candidate_id, features) in batch.cpu_rows.drain(..) {
        let input = &inputs[candidate_id];
        let text = input.ml_text();
        let features = features.unwrap_or_else(|| input.ml_features(config));
        batch.scores[candidate_id] =
            crate::confidence::policy::ml_score_for_candidate_text(text, || {
                forward_pass(&features) as f64
            });
    }
}

#[cfg(feature = "ml")]
pub(crate) fn score_input_batch_quantized_cpu<T: MlScoreInput>(
    inputs: &[T],
    config: &crate::types::ScannerConfig,
) -> crate::Result<Vec<f64>> {
    let mut batch = prepare_quantized_batch(inputs, config)
        .map_err(|error| crate::ScanError::Config(error.to_string()))?;
    let accelerated_scores = crate::confidence::quantized::score_batch(&batch.accelerated_rows)
        .map_err(|error| crate::ScanError::Config(error.to_string()))?;
    score_cpu_owned_rows(&mut batch, inputs, config);
    for (candidate_id, score) in batch
        .accelerated_candidate_ids
        .into_iter()
        .zip(accelerated_scores)
    {
        batch.scores[candidate_id] = score.as_f64();
    }
    Ok(batch.scores)
}

/// Score GPU-owned candidates on the selected VYRE peer and keep explicitly
/// ineligible candidates on the established CPU model.
#[cfg(all(feature = "gpu", feature = "ml"))]
pub(crate) fn score_input_batch_quantized_vyre<T: MlScoreInput>(
    inputs: &[T],
    config: &crate::types::ScannerConfig,
    backend: &dyn vyre::VyreBackend,
    deadline: Option<std::time::Instant>,
) -> crate::Result<Vec<f64>> {
    let mut batch = prepare_quantized_batch(inputs, config)
        .map_err(|error| crate::ScanError::Gpu(error.to_string()))?;
    let timeout =
        deadline.map(|deadline| deadline.saturating_duration_since(std::time::Instant::now()));
    let pending_gpu_scores =
        crate::confidence::quantized_vyre::submit_rows(backend, &batch.accelerated_rows, timeout)
            .map_err(|error| crate::ScanError::Gpu(error.to_string()))?;
    score_cpu_owned_rows(&mut batch, inputs, config);
    let gpu_scores = crate::confidence::quantized::validate_accelerated_output(
        batch.accelerated_rows.len(),
        pending_gpu_scores.await_scores(),
    )
    .map_err(|error| crate::ScanError::Gpu(error.to_string()))?;
    for (candidate_id, score) in batch.accelerated_candidate_ids.into_iter().zip(gpu_scores) {
        batch.scores[candidate_id] = score.as_f64();
    }
    Ok(batch.scores)
}

/// Score precomputed model features without recomputing text/context signals.
#[cfg(feature = "ml")]
pub(crate) fn score_features(features: &[f32; NUM_FEATURES]) -> f64 {
    forward_pass(features) as f64
}

/// Score an already-extracted batch on CPU while acquiring the immutable model
/// once for the whole batch.
#[cfg(feature = "ml")]
pub(crate) fn score_precomputed_batch_on_cpu<T: MlScoreInput>(
    inputs: &[T],
    features: &[[f32; NUM_FEATURES]],
) -> Vec<f64> {
    use rayon::prelude::*;

    assert_eq!(
        inputs.len(),
        features.len(),
        "internal invariant violation: ML input and feature batch lengths differ"
    );
    let model = ml_weights::model();
    inputs
        .par_iter()
        .zip(features.par_iter())
        .map(|(input, features)| {
            crate::confidence::policy::ml_score_for_candidate_text(input.ml_text(), || {
                forward_pass_impl(model, features) as f64
            })
        })
        .collect()
}

/// Fused serial feature extraction and CPU scoring for the common tiny batch.
/// The model pointer is resolved once while feature arrays remain stack-local.
#[cfg(feature = "ml")]
pub(crate) fn score_input_batch_serial<T: MlScoreInput>(
    inputs: &[T],
    config: &crate::types::ScannerConfig,
) -> Vec<f64> {
    let model = ml_weights::model();
    inputs
        .iter()
        .map(|input| {
            crate::confidence::policy::ml_score_for_candidate_text(input.ml_text(), || {
                forward_pass_impl(model, &input.ml_features(config)) as f64
            })
        })
        .collect()
}

/// Score one complete pending-match batch without transferring confidence
/// policy to an accelerator. GPU routes accelerate VYRE-owned detection work;
/// final confidence scoring remains the same deterministic CPU implementation
/// for every backend.
#[cfg(feature = "ml")]
pub(crate) fn score_input_batch<T: MlScoreInput>(
    inputs: &[T],
    config: &crate::types::ScannerConfig,
) -> Vec<f64> {
    if inputs.is_empty() {
        return Vec::new();
    }

    let profile = keyhog_profile::enabled();
    if inputs.len() < ML_PARALLEL_BATCH_THRESHOLD {
        if !profile {
            return score_input_batch_serial(inputs, config);
        }

        let model = ml_weights::model();
        let mut feature_ns = 0_u64;
        let mut score_ns = 0_u64;
        let mut scores = Vec::with_capacity(inputs.len());
        for input in inputs {
            let feature_started = std::time::Instant::now();
            let features = input.ml_features(config);
            feature_ns = feature_ns.saturating_add(feature_started.elapsed().as_nanos() as u64);

            let score_started = std::time::Instant::now();
            scores.push(crate::confidence::policy::ml_score_for_candidate_text(
                input.ml_text(),
                || forward_pass_impl(model, &features) as f64,
            ));
            score_ns = score_ns.saturating_add(score_started.elapsed().as_nanos() as u64);
        }
        keyhog_profile::add_counter(keyhog_profile::CounterId::MlFeatureNs, feature_ns);
        keyhog_profile::add_counter(keyhog_profile::CounterId::MlScoreNs, score_ns);
        return scores;
    }

    use rayon::prelude::*;
    if !profile {
        let model = ml_weights::model();
        return inputs
            .par_iter()
            .map(|input| {
                let features = input.ml_features(config);
                crate::confidence::policy::ml_score_for_candidate_text(input.ml_text(), || {
                    forward_pass_impl(model, &features) as f64
                })
            })
            .collect();
    }

    let feature_started = std::time::Instant::now();
    let features: Vec<[f32; NUM_FEATURES]> = inputs
        .par_iter()
        .map(|input| input.ml_features(config))
        .collect();
    keyhog_profile::add_counter(
        keyhog_profile::CounterId::MlFeatureNs,
        feature_started.elapsed().as_nanos() as u64,
    );

    let score_started = std::time::Instant::now();
    let scores = score_precomputed_batch_on_cpu(inputs, &features);
    keyhog_profile::add_counter(
        keyhog_profile::CounterId::MlScoreNs,
        score_started.elapsed().as_nanos() as u64,
    );
    scores
}

/// Return the embedded model version string for diagnostics and CLI output.
pub fn model_version() -> &'static str {
    ml_weights::MODEL_VERSION
}

/// Return a compact embedded model-card summary for diagnostics and CLI output.
pub fn model_card_summary() -> &'static str {
    ml_weights::MODEL_CARD_SUMMARY
}

/// Return the full embedded model-card JSON for provenance-aware tooling.
pub fn model_card_json() -> &'static str {
    ml_weights::MODEL_CARD_JSON
}

/// Forward pass through the MoE model with hardcoded weights.
///
/// Two layered optimizations, both numerically inert:
///
/// 1. **Weight hoist.** The model's weight/bias slices are resolved ONCE per call
///    via [`ml_weights::model`] (a single `OnceLock`-acquire of an already-built
///    `&'static MoeModel`) instead of the prior 37 per-candidate accessor calls,
///    each of which re-acquired the `OnceLock` and re-sliced the flat buffer.
///
/// 2. **Output-stationary dense layers.** The two large expert layers (fc1: 32
///    outputs, fc2: 16) run via [`dense_relu_layer_t`] over COLUMN-major
///    (transposed) weights: for each input the contiguous output row is scaled
///    and accumulated, so the dependency-free inner loop over outputs vectorizes
///    across 8-16 output lanes. Each output still reduces its inputs in index
///    order with separate round(mul)/round(add) (no FMA fusion), so the result is
///    BIT-IDENTICAL to the row-major scalar dot product, vectorizing across
///    outputs never reassociates a single output's sum. The small gate (6) and
///    fc3 (1) layers stay scalar [`dense_row`] (nothing to vectorize across).
///
/// An explicit AVX2+FMA reduction was tried INSTEAD of (2) and REVERTED:
/// `_mm256_fmadd_ps` fuses each multiply-add with a single rounding step and
/// reassociates the sum across 8 lanes, so its result is NOT bit-identical. That
/// sub-ULP divergence pushed borderline ML-gated detectors (twilio-auth-token,
/// africastalking-api-key, appsmith-api-credentials, …) across their
/// `min_confidence` floor and regressed 30+ `contracts_runner` positives/evasions.
/// The confidence model and the GPU parity reference (DET-11) are calibrated
/// against this exact reduction. The output-stationary layout gets the SIMD width
/// WITHOUT the divergence (proved bit-identical in
/// `tests/ml_forward_parity.rs`); do NOT reintroduce FMA fusion or lane
/// reassociation without recalibrating every contract and the GPU parity reference.
fn forward_pass(input: &[f32; NUM_FEATURES]) -> f32 {
    let model = ml_weights::model();
    forward_pass_impl(model, input)
}

/// MoE forward pass over the hoisted model. fc1/fc2 use the output-stationary
/// vectorized kernel ([`dense_relu_layer_t`]); the gate and fc3 stay scalar.
fn forward_pass_impl(model: &ml_weights::MoeModel, input: &[f32; NUM_FEATURES]) -> f32 {
    let gate_probs = softmax(&compute_gate_logits(model, input));
    let mut score_logit = 0.0f32;
    for (expert_idx, gate_prob) in gate_probs.iter().enumerate() {
        score_logit += *gate_prob * expert_logit(&model.experts[expert_idx], input);
    }
    sigmoid(score_logit)
}

fn compute_gate_logits(
    model: &ml_weights::MoeModel,
    input: &[f32; NUM_FEATURES],
) -> [f32; EXPERT_COUNT] {
    debug_assert_eq!(model.gate_weight.len(), NUM_FEATURES * EXPERT_COUNT);
    debug_assert_eq!(model.gate_bias.len(), EXPERT_COUNT);

    let mut gate_logits = [0.0f32; EXPERT_COUNT];
    for (expert_idx, logit) in gate_logits.iter_mut().enumerate() {
        let row = &model.gate_weight[expert_idx * NUM_FEATURES..];
        *logit = dense_row(row, input, model.gate_bias[expert_idx]);
    }
    gate_logits
}

fn expert_logit(expert: &ml_weights::ExpertWeights, input: &[f32; NUM_FEATURES]) -> f32 {
    let h1 = dense_relu_layer_t::<NUM_FEATURES, EXPERT_FC1_OUT>(
        expert.fc1_weight_t,
        expert.fc1_bias,
        input,
    );
    let h2 = dense_relu_layer_t::<EXPERT_FC1_OUT, EXPERT_FC2_OUT>(
        expert.fc2_weight_t,
        expert.fc2_bias,
        &h1,
    );
    dense_row(expert.fc3_weight, &h2, expert.fc3_bias)
}

/// Output-stationary ReLU dense layer over COLUMN-major (transposed) weights:
/// `weights_t[k*OUTPUT + o]` is input `k`'s weight to output `o`
/// (`ml_weights::transpose_static`). For each input `k` we scale its contiguous
/// `OUTPUT`-wide weight row by `input[k]` and accumulate into the `OUTPUT` running
/// sums. The inner loop over outputs has NO loop-carried dependency (each
/// `acc[o]` is independent), so LLVM vectorizes it across the output lanes at
/// opt-level 3: 8-16 outputs updated per SIMD instruction instead of one scalar
/// MAC at a time.
///
/// BIT-IDENTICAL to the row-major `dense_row` dot product: each `acc[o]` still
/// starts at `bias[o]` and adds `input[k]*w[o][k]` for `k = 0,1,..,INPUT-1` in
/// order, with a separate round-to-f32 on the multiply and on the add (Rust does
/// NOT contract `a*b + c` into a fused multiply-add without fast-math, and we use
/// none), then the SAME `.max(0.0)` ReLU. Vectorizing ACROSS outputs does not
/// reassociate any single output's reduction, so the result equals the scalar
/// path bit-for-bit. The previous AVX2+FMA attempt reassociated lanes and fused
/// the MAC, diverged sub-ULP, and regressed ~30 ML-gated contracts; this layout
/// does not. The equality is proved exhaustively over random weights/inputs in
/// `crates/scanner/tests/ml_forward_parity.rs`.
#[inline]
fn dense_relu_layer_t<const INPUT: usize, const OUTPUT: usize>(
    weights_t: &[f32],
    bias: &[f32],
    input: &[f32; INPUT],
) -> [f32; OUTPUT] {
    let mut acc = [0.0f32; OUTPUT];
    for (o, slot) in acc.iter_mut().enumerate() {
        *slot = bias[o];
    }
    for k in 0..INPUT {
        let x = input[k];
        // One contiguous OUTPUT-wide weight row per input. `zip` bounds the
        // iteration to `min(OUTPUT, row.len())` with no per-element bounds check,
        // and vectorizes across the output lanes.
        let row = &weights_t[k * OUTPUT..k * OUTPUT + OUTPUT];
        for (slot, &w) in acc.iter_mut().zip(row.iter()) {
            *slot += x * w;
        }
    }
    for slot in acc.iter_mut() {
        *slot = slot.max(0.0);
    }
    acc
}

/// Dot product of a weight row with the input vector plus bias.
///
/// `weights` may be longer than `input` (it is a borrow into the flat model
/// buffer starting at the row offset); `zip` bounds the reduction to exactly
/// `INPUT` pairs with no per-element bounds check, and the statically-known
/// `input` length lets the backend autovectorize. The accumulation stays a
/// single left-to-right sequential sum (`i = 0,1,..,INPUT-1`) with a separate
/// round on the multiply and the add (no FMA fusion), so the f32 result is
/// bit-identical to the scalar reference.
#[inline(always)]
fn dense_row<const INPUT: usize>(weights: &[f32], input: &[f32; INPUT], bias: f32) -> f32 {
    let mut sum = bias;
    for (&x, &w) in input.iter().zip(weights.iter()) {
        sum += x * w;
    }
    sum
}

pub(crate) fn sigmoid(value: f32) -> f32 {
    let x = value;
    if x <= -SIGMOID_SATURATION {
        0.0
    } else if x >= SIGMOID_SATURATION {
        1.0
    } else {
        // Fast polynomial/rational evaluation of sigmoid (0.5 + 0.5 * x / (1 + |x|))
        // which avoids expensive transcendental exp() function calls.
        0.5 + 0.5 * x / (1.0 + x.abs())
    }
}

fn softmax(logits: &[f32; EXPERT_COUNT]) -> [f32; EXPERT_COUNT] {
    let max_logit = logits.iter().copied().fold(f32::NEG_INFINITY, f32::max);
    let mut exps = [0.0f32; EXPERT_COUNT];
    let mut sum = 0.0f32;
    for (idx, logit) in logits.iter().enumerate() {
        let value = (*logit - max_logit).exp();
        exps[idx] = value;
        sum += value;
    }
    for value in &mut exps {
        *value /= sum;
    }
    exps
}