ferrox_models/

loader.rs

//! Loads a real `Decoder` from an on-disk GGUF file, using the
//! llama.cpp-style tensor naming convention
//! (`token_embd.weight`, `blk.N.attn_q.weight`, `blk.N.ffn_gate.weight`
//! or, for MoE, `blk.N.ffn_gate_exps.weight`, `output_norm.weight`,
//! `output.weight`). Until this module existed, ferrox could only run
//! correctly-shaped *random* weights.
//!
//! Quantized tensors (Q8_0 / Q4_0) are loaded as `WeightMatrix::Quantized`
//! backed by `WeightBytes::Mapped` -- a zero-copy view into the same
//! mmap `GgufFile` already holds, with no intermediate heap copy of the
//! tensor's bytes at all. So a checkpoint's resident memory is the
//! mmap page cache, not the mmap plus a second in-process copy of every
//! weight. `WeightMatrix::apply` dispatches to ferrox-quant's fused
//! dequant+dot kernels directly against those mapped bytes at inference
//! time. F32 tensors (norms, embeddings, and any weight not natively
//! quantized) still copy into an owned `Tensor`, since they're small
//! relative to the quantized weight matrices and need per-element
//! access patterns a raw byte view doesn't support as cleanly.
//!
//! Verified end to end (see `crates/ferrox-models/tests/gguf_roundtrip.rs`)
//! against a genuinely Q8_0-quantized, generated on-disk GGUF fixture
//! for the dense (single-expert) case, and against real OLMoE / Qwen2-MoE
//! checkpoints for the multi-expert 3D-packed-tensor path.

use ferrox_core::expert_store::{ExpertKey, ExpertSource, ExpertStore};
use ferrox_core::tensor::Tensor;
use ferrox_core::weight_matrix::quant_kind_for;
use ferrox_core::weight_matrix::{QuantKind, WeightBytes, WeightMatrix};
use ferrox_gguf::{GgmlType, GgufError, GgufValue, ShardedGguf, TensorInfo, TensorSource};
use ferrox_moe::{ExpertWeights, GatingFunction, MoeLayerConfig};
use std::sync::Arc;
use thiserror::Error;

use crate::config::ModelConfig;
#[cfg(feature = "metal")]
use crate::decoder::MoePackedQ4Planes;
use crate::decoder::{AttnWeights, Decoder, ExpertBacking, LayerWeights, MoeWeights};

#[derive(Debug, Error)]
pub enum LoadError {
    #[error(transparent)]
    Gguf(#[from] GgufError),
    #[error(transparent)]
    Shard(#[from] ferrox_gguf::ShardError),
    #[error("tensor '{0}' has unsupported dtype {1:?}")]
    UnsupportedDtype(String, GgmlType),
    #[error(
        "MoE tensor '{0}' is not 3D or its expert count {1} does not match config n_experts {2}"
    )]
    ExpertCountMismatch(String, usize, usize),
    #[error("GGUF file is missing required hparam metadata key '{0}'")]
    MissingHparam(String),
    /// `general.architecture` is not in the capability registry -- refuse
    /// to guess RoPE/gating rather than emit fluent-but-wrong logits.
    #[error(
        "unsupported GGUF architecture '{0}': not in ferrox's capability registry \
         (unknown required features fail closed; see ferrox_models::capability)"
    )]
    UnsupportedArchitecture(String),
    /// Architecture exists but must not use the generic GQA decoder.
    #[error("architecture '{0}' cannot use the generic Decoder: {1}")]
    DedicatedArchitectureRequired(String, &'static str),
    /// Metadata advertises a feature the generic decoder does not implement.
    #[error("architecture '{0}' requires unimplemented feature: {1}")]
    UnsupportedFeature(String, String),
    #[error(
        "architecture '{0}' has never been verified against llama.cpp. It would run on \
         ferrox's shared generic-GQA path, which ASSUMES plain GQA with {1:?} RoPE and no \
         ALiBi, no learned position embeddings and no per-layer rope skipping. That \
         assumption has already been wrong for gpt2, mpt, refact, bloom and jais, each of \
         which loaded clean and answered as a different model. {2} Set \
         FERROX_ALLOW_UNAUDITED_ARCH=1 to run it anyway and compare the output against \
         llama.cpp yourself"
    )]
    UnauditedArchitecture(String, crate::config::RopeLayout, String),
    /// The checkpoint carries per-block tensors this build never reads,
    /// i.e. weights that contribute to the real graph and would simply
    /// be missing from ours. See [`assert_every_tensor_consumed`].
    #[error(
        "checkpoint carries {0} tensor(s) this build never reads, so its graph is not the one \
         ferrox would run: {1}. This is a missing feature, not a corrupt file. Override with \
         FERROX_ALLOW_UNKNOWN_TENSORS=1 to load anyway and accept wrong output."
    )]
    UnconsumedTensors(usize, String),
    /// `FERROX_STRICT_KERNELS=1` and the model has weights with no
    /// kernel on the selected accelerator, i.e. it would run, correctly,
    /// on a silently slower path. Refusing is the point: a benchmark or
    /// CI run must not be able to publish a number taken off the
    /// backend it claims. See [`ferrox_core::kernel_registry`].
    #[error("{0}")]
    StrictKernels(String),
}

/// Architecture-family name strings (GGUF's `general.architecture` value)
/// known, from reading ik_llama.cpp's `llama-hparams.cpp`
/// (`LLM_ARCH_DEEPSEEK2`, `LLM_ARCH_GLM4_MOE` cases), to default to
/// sigmoid MoE gating with post-selection renormalization rather than
/// softmax. Every member's citation is inline here; `docs/MODELS.md`
/// carries none and the pointer that used to send readers there was
/// dangling.
/// `afmoe`, `laguna` and `step35` added 2026-09-01 by the
/// unaudited-refusal triage's gating sweep. Each reads
/// `LLM_KV_EXPERT_GATING_FUNC` as OPTIONAL and then, when the key is
/// absent, sets `LLAMA_EXPERT_GATING_FUNC_TYPE_SIGMOID`
/// (`afmoe.cpp:29-30`, `laguna.cpp:55-56`, `step35.cpp:19-20`). Ferrox
/// fell back to softmax for all three.
///
/// This is the `deepseek` shape a third, fourth and fifth time: a
/// default that is right for most architectures and silently wrong for
/// one, where the GGUF carries no key to correct it. Nothing is live
/// today -- all three are `NewCode` for other reasons and refuse before
/// reaching here -- but the list is what a later admission would trust.
const SIGMOID_GATING_ARCHITECTURES: &[&str] =
    &["afmoe", "deepseek2", "glm4moe", "laguna", "step35"];

/// Names that appear in a behaviour table above but are `DedicatedOnly`
/// or `Deferred`, together with the module that actually applies the
/// behaviour for them.
///
/// Two true things were in conflict here, and deleting either would
/// have lost one. `SIGMOID_GATING_ARCHITECTURES` records a fact about
/// llama.cpp (these architectures default to sigmoid when the GGUF
/// carries no `expert_gating_func`), and a test pins it as such. The
/// cross-table test records a different fact: an entry for an
/// architecture that never reaches THIS loader cannot fire, and a gate
/// that cannot fire is worse than no gate because it reads as coverage.
///
/// Both hold. `deepseek2` and `glm4moe` are genuinely sigmoid-gated and
/// genuinely never arrive here. So the resolution is not to drop a name
/// from either place, it is to say out loud who owns it instead, and to
/// make an unexplained dead entry still fail.
///
/// Adding a name here is a claim that the named module applies the
/// behaviour. It is checked no further than that, so it is the one line
/// in this file to be suspicious of.
/// Test-only: it asserts a relationship rather than driving one, and a
/// production reader would have to be told that.
#[cfg(test)]
const DEDICATED_OWNS_ITS_BEHAVIOUR: &[(&str, &str)] = &[
    // `mla_gguf_loader` reads `expert_gating_func` and falls back to
    // Sigmoid itself, so deepseek2's gating is decided there.
    ("deepseek2", "mla_gguf_loader"),
    // glm4moe is refused today (it needs gpt-oss's norm slot, see its
    // refusal text). The entry stays because the fact about llama.cpp
    // stays true, and it becomes live the moment the refusal lifts.
    ("glm4moe", "refused today, see capability::unaudited_triage"),
];

/// Architecture-family names whose real reference implementation skips
/// renormalizing top-k softmax routing weights after selection (GGUF
/// carries no metadata key for this -- it's hardcoded per-architecture in
/// both the real HF `transformers` model code and llama.cpp's
/// `build_moe_ffn` call sites, not read from the file). Confirmed for
/// `olmoe` against `OlmoeTopKRouter.forward` in
/// `transformers/models/olmoe/modeling_olmoe.py` (`config.norm_topk_prob`
/// is `false` in the real published config.json) and llama.cpp's
/// `src/models/olmoe.cpp` (`build_moe_ffn(..., false, ...,
/// LLAMA_EXPERT_GATING_FUNC_TYPE_SOFTMAX, ...)`). See
/// `MoeLayerConfig::norm_topk_prob`'s doc comment for why this matters:
/// getting it wrong silently produces wrong generation output even
/// though the file loads and shape-validates fine.
// Architectures whose reference graphs pass `norm_w=false` to
// `build_moe_ffn` (llama.cpp) / `norm_topk_prob=false` in HF config.
// Qwen2-MoE: `.scratch/llama.cpp/src/models/qwen2moe.cpp` -- Softmax +
// `false` for the norm_topk slot. Renormalizing top-k weights made
// Qwen1.5-MoE greedy decode emit garbage despite shared-expert load.
// `deepseek` (V1) added 2026-09-01 by the unaudited-refusal triage.
// `src/models/deepseek.cpp:145-155` passes `norm_w=false`, and
// `conversion/deepseek.py`'s `DeepseekModel` never writes
// `{arch}.expert_weights_norm` -- only `DeepseekV2Model` does -- so no
// real `deepseek` GGUF carries the key to override the default with.
// Ferrox therefore renormalised where llama.cpp does not. Same class of
// bug as the OLMoE one above, and latent only because `deepseek` is
// unaudited and refuses first.
const NO_TOPK_RENORMALIZE_ARCHITECTURES: &[&str] = &["deepseek", "olmoe", "qwen2moe"];

/// Architectures that store their **pre-FFN** norm under the tensor name
/// `blk.N.post_attention_norm.weight` and carry no `blk.N.ffn_norm`.
///
/// Gemma writes the same tensor name for a genuinely different norm: it
/// is applied to the attention output *inside* the attention residual,
/// and Gemma also carries `ffn_norm`. Reading one file's tensor with the
/// other's meaning silently moves a whole RMSNorm to the wrong side of a
/// residual add, so the meaning is decided by architecture, not by which
/// tensors happen to be present.
///
/// - `gpt-oss`: `openai-moe.cpp` norms `ffn_inp` with `attn_post_norm`.
/// - `seed_oss`: `src/models/seed-oss.cpp:36-37` creates `attn_norm` and
///   `attn_post_norm` and **no** `ffn_norm`, and `:113-115` norms
///   `ffn_inp` -- the post-attention residual -- with `attn_post_norm`.
///
/// This is deliberately NOT `arch == "gpt-oss"`, which is what it used
/// to be. That one flag also gated gpt-oss's five extra per-layer
/// tensors (sinks, biases, the SwiGLU clamp), and widening it would have
/// handed `seed_oss` attention sinks it does not have. Two facts, two
/// predicates. `the_norm_slot_list_and_the_audit_list_agree` pins that a
/// name added here is a name somebody actually read a graph for.
const PRE_FFN_NORM_IS_POST_ATTENTION_NORM: &[&str] = &["gpt-oss", "seed_oss"];

/// Does this architecture keep its pre-FFN norm in the
/// `post_attention_norm` slot? See
/// [`PRE_FFN_NORM_IS_POST_ATTENTION_NORM`].
fn pre_ffn_norm_is_post_attention_norm(arch: &str) -> bool {
    PRE_FFN_NORM_IS_POST_ATTENTION_NORM.contains(&arch)
}

/// Architectures whose checkpoints carry `{arch}.leading_dense_block_count`
/// while their reference graph never branches on it: **every** layer is
/// MoE regardless of what the key says.
///
/// `bailingmoe` is the case this list exists for.
/// `src/models/bailingmoe.cpp:5` reads
/// `LLM_KV_LEADING_DENSE_BLOCK_COUNT` into `n_layer_dense_lead` and then
/// `load_arch_tensors` creates `ffn_gate_inp`, the expert tensors and
/// the shared-expert tensors unconditionally for every layer (:39-54 --
/// there is no `if (i < n_layer_dense_lead)` anywhere in the file) and
/// the graph has no dense branch either (:119-152). Meanwhile
/// `conversion/bailingmoe.py:27` writes `first_k_dense_replace` into the
/// key verbatim, so real Ling checkpoints DO carry a nonzero value.
///
/// Ferrox's `ModelConfig::layer_is_dense` does branch on it, so without
/// this list ferrox looks for `blk.0.ffn_gate.weight` on a layer that
/// only ships experts and dies on a missing tensor. That is a load
/// failure rather than wrong logits, which is why it stayed latent.
///
/// Do not read this as "the key is meaningless": for `deepseek`,
/// `dots1`, `glm4moe` and every other leading-dense architecture the key
/// is load-bearing and must be honoured. Membership here is a statement
/// about ONE architecture's graph, checked in that graph.
const LEADING_DENSE_KEY_IS_INERT: &[&str] = &["bailingmoe"];

/// Architectures whose reference graph applies `attn_q_norm` /
/// `attn_k_norm` AFTER `ggml_rope_ext`, not before it.
///
/// There is no GGUF key for this. llama.cpp writes the order into each
/// hand-written graph, so the only place it can come from is the
/// architecture string, and getting it wrong changes every layer's
/// attention scores without changing a single tensor shape.
///
/// - `maincoder`: `src/models/maincoder.cpp:78-90` ropes Q and K, then
///   norms them at `:92` and `:95`.
/// - `hunyuan-moe`: `src/models/hunyuan-moe.cpp:93,104` rope, `:110,115`
///   norm.
///
/// - `hunyuan-dense`: it has no graph of its own --
///   `src/models/models.h:1830-1834` derives `llama_model_hunyuan_dense`
///   from `llama_model_hunyuan_vl` and reuses its graph -- so the lines
///   are `hunyuan-vl.cpp:56-66` (rope) then `:73-81` (norm). Its second
///   blocker, the NTK-alpha RoPE base rescale, is implemented too; see
///   [`crate::rope_ntk_alpha`].
///
/// The audited majority is the other way round -- `qwen3moe.cpp:99,108`
/// and `bailingmoe2.cpp:123-135` both norm first -- which is why the
/// decoder's default is "before" and this list is the exception.
const QK_NORM_AFTER_ROPE_ARCHITECTURES: &[&str] = &["hunyuan-dense", "hunyuan-moe", "maincoder"];

fn metadata_u64_any(file: &impl TensorSource, keys: &[String]) -> Option<u64> {
    keys.iter().find_map(|k| file.metadata_u64(k))
}

fn metadata_f32_any(file: &impl TensorSource, keys: &[String]) -> Option<f32> {
    keys.iter()
        .find_map(|k| file.metadata(k).and_then(GgufValue::as_f32))
}

impl ModelConfig {
    /// Derives a `ModelConfig` from a real GGUF file's own hyperparameter
    /// metadata, following llama.cpp's `general.architecture`-prefixed key
    /// convention (`{arch}.block_count`, `{arch}.embedding_length`,
    /// `{arch}.attention.head_count`, `{arch}.expert_count`, ...) rather
    /// than requiring a hand-written preset to already match the file's
    /// shape exactly. This is what lets `ferrox-server` (and `ferrox
    /// run-real`) load an arbitrary checkpoint, not just the three
    /// hand-tuned presets in `config.rs`.
    ///
    /// Fields with no corresponding metadata key fall back to widely-used
    /// llama.cpp defaults (documented inline) and are listed in the
    /// returned config's `best_effort_fields`, following the same
    /// confirmed-vs-estimated discipline as the hand-written presets.
    pub fn from_gguf(file: &impl TensorSource) -> Result<Self, LoadError> {
        let arch = file
            .metadata_str("general.architecture")
            .ok_or_else(|| LoadError::MissingHparam("general.architecture".to_string()))?
            .to_string();
        let arch_profile = crate::capability::resolve_profile(&arch)
            .ok_or_else(|| LoadError::UnsupportedArchitecture(arch.clone()))?;
        let rope_layout = match arch_profile.path {
            crate::capability::ArchPath::GenericGqa { rope }
            | crate::capability::ArchPath::TestFixture { rope } => rope,
            crate::capability::ArchPath::DedicatedOnly { reason } => {
                return Err(LoadError::DedicatedArchitectureRequired(
                    arch.clone(),
                    reason,
                ));
            }
            crate::capability::ArchPath::Deferred { reason } => {
                return Err(LoadError::UnsupportedFeature(
                    arch.clone(),
                    format!("architecture deferred from Ferrox text-generation scope: {reason}"),
                ));
            }
        };
        let qk_norm_style = arch_profile.qk_norm;
        for (meta_key, feature) in crate::capability::unsupported_feature_keys(&arch) {
            if let Some(v) = metadata_f32_any(file, std::slice::from_ref(&meta_key)) {
                if v > 0.0 {
                    return Err(LoadError::UnsupportedFeature(
                        arch.clone(),
                        format!("{feature} (metadata {meta_key}={v})"),
                    ));
                }
            }
            if let Some(v) = metadata_u64_any(file, std::slice::from_ref(&meta_key)) {
                if v > 0 {
                    return Err(LoadError::UnsupportedFeature(
                        arch.clone(),
                        feature.to_string(),
                    ));
                }
            }
        }
        // Metadata-declared multipliers the generic decoder does not
        // apply. Unlike the tensor-consumption gate, nothing about these
        // is visible in the weights, so a Granite checkpoint would load
        // and answer at the wrong scale. See
        // `capability::unsupported_scaling_keys`.
        for (meta_key, feature, no_op) in crate::capability::unsupported_scaling_keys(&arch) {
            if let Some(v) = metadata_f32_any(file, std::slice::from_ref(&meta_key)) {
                if (v - no_op).abs() > 1e-6 {
                    return Err(LoadError::UnsupportedFeature(
                        arch.clone(),
                        format!("{feature} (metadata {meta_key}={v})"),
                    ));
                }
            }
        }
        let key = |suffix: &str| format!("{arch}.{suffix}");

        let name: &'static str = Box::leak(
            file.metadata_str("general.name")
                .unwrap_or(&arch)
                .to_string()
                .into_boxed_str(),
        );

        let n_layers =
            file.metadata_u64(&key("block_count"))
                .ok_or_else(|| LoadError::MissingHparam(key("block_count")))? as usize;
        // Baichuan is one architecture string covering two positional
        // schemes: 7B rotates, 13B uses ALiBi and no RoPE at all
        // (`src/models/baichuan.cpp:11-14`, `:57-58`, where `inp_pos` is
        // `nullptr` for 13B, so `ggml_rope_ext` is never reached).
        // llama.cpp decides that on the layer count and says so in a
        // comment: "TODO: become GGUF KV parameter". There is therefore
        // no key for `capability::unsupported_feature_keys` to test and
        // no tensor for `assert_every_tensor_consumed` to miss. A
        // Baichuan-13B checkpoint loads clean and is rotated anyway.
        // Refuse it here, where the layer count is known.
        if arch == "baichuan" && n_layers == 40 {
            return Err(LoadError::UnsupportedFeature(
                arch.clone(),
                "Baichuan-13B (block_count=40) uses ALiBi and no RoPE, decided by layer \
                 count with no GGUF key to declare it; the generic decoder would rotate \
                 every Q/K head instead. Baichuan-7B (block_count=32) is unaffected"
                    .to_string(),
            ));
        }
        let hidden_dim = file
            .metadata_u64(&key("embedding_length"))
            .ok_or_else(|| LoadError::MissingHparam(key("embedding_length")))?
            as usize;
        let n_heads = file
            .metadata_u64(&key("attention.head_count"))
            .ok_or_else(|| LoadError::MissingHparam(key("attention.head_count")))?
            as usize;

        let mut best_effort_fields: Vec<&'static str> = Vec::new();

        let n_kv_heads = file
            .metadata_u64(&key("attention.head_count_kv"))
            .map(|v| v as usize)
            .unwrap_or_else(|| {
                best_effort_fields.push("n_kv_heads (no attention.head_count_kv key; assumed equal to n_heads, i.e. plain MHA)");
                n_heads
            });
        let head_dim = file
            .metadata_u64(&key("attention.key_length"))
            .map(|v| v as usize)
            .unwrap_or_else(|| {
                best_effort_fields.push(
                    "head_dim (no attention.key_length key; derived as hidden_dim / n_heads)",
                );
                hidden_dim / n_heads
            });
        let v_head_dim = file
            .metadata_u64(&key("attention.value_length"))
            .map(|v| v as usize)
            .unwrap_or(head_dim);
        if v_head_dim != head_dim {
            return Err(LoadError::UnsupportedFeature(
                arch.clone(),
                format!(
                    "split K/V head dims (key_length={head_dim}, value_length={v_head_dim}); \
                     generic decoder requires equal head dims"
                ),
            ));
        }
        let vocab_size = file
            .metadata("tokenizer.ggml.tokens")
            .and_then(|v| match v {
                GgufValue::Array(items) => Some(items.len()),
                _ => None,
            })
            .or_else(|| file.metadata_u64(&key("vocab_size")).map(|v| v as usize))
            .unwrap_or_else(|| {
                best_effort_fields.push("vocab_size (no tokenizer.ggml.tokens array or {arch}.vocab_size key; fell back to output.weight's own row count)");
                // `output.weight`'s real raw shape is `[hidden_dim,
                // vocab_size]` (ggml's fastest-first `ne[]` order --
                // see `load_weight_matrix`'s doc comment), so vocab_size
                // is the *last* element, not the first.
                file.find_tensor("output.weight")
                    .and_then(|t| t.shape.last().copied())
                    .unwrap_or(0) as usize
            });
        let rope_theta = metadata_f32_any(file, &[key("rope.freq_base")]).unwrap_or_else(|| {
            best_effort_fields.push("rope_theta (no rope.freq_base key; defaulted to 10000.0)");
            10000.0
        });
        // NTK-alpha: `{arch}.rope.scaling.alpha` is read for every
        // architecture (llama-model.cpp:1186) and APPLIED by two
        // (`hunyuan-vl.cpp:8-12`, inherited by `hunyuan-dense`). The
        // list and the arithmetic live together in one module so the
        // key's readers and its appliers cannot drift apart -- see
        // `crate::rope_ntk_alpha`, which also records why a converted
        // `hunyuan-dense` file carries the already-scaled base instead.
        let rope_theta = crate::rope_ntk_alpha::ntk_alpha_scaled_rope_base(
            &arch,
            rope_theta,
            head_dim,
            metadata_f32_any(file, &[key("rope.scaling.alpha")]),
        );
        let rms_norm_eps = metadata_f32_any(
            file,
            &[
                key("attention.layer_norm_rms_epsilon"),
                key("attention.layer_norm_epsilon"),
            ],
        )
        .unwrap_or_else(|| {
            best_effort_fields
                .push("rms_norm_eps (no layer_norm_rms_epsilon key; defaulted to 1e-5)");
            1e-5
        });

        let n_experts = metadata_u64_any(file, &[key("expert_count")]).unwrap_or(0) as usize;
        let is_moe = n_experts > 1;

        let n_experts_active = if is_moe {
            metadata_u64_any(file, &[key("expert_used_count")]).unwrap_or_else(|| {
                best_effort_fields
                    .push("moe.n_experts_active (no expert_used_count key; defaulted to 2)");
                2
            }) as usize
        } else {
            1
        };
        // Prefer the GGUF hparam when present. Qwen2MoE (and some other
        // HF→GGUF exports) omit `expert_shared_count` but still ship
        // `blk.N.ffn_{gate,up,down}_shexp.weight` -- without a tensor-
        // presence fallback those weights are silently dropped and the
        // model runs with a large chunk of active FFN missing.
        let n_shared_experts = match metadata_u64_any(file, &[key("expert_shared_count")]) {
            Some(n) => n as usize,
            None if is_moe && file.find_tensor("blk.0.ffn_gate_shexp.weight").is_some() => {
                best_effort_fields.push(
                    "moe.n_shared_experts (no expert_shared_count; inferred 1 from blk.0.ffn_gate_shexp.weight)",
                );
                1
            }
            None => 0,
        };
        // MoE GGUFs often only set `feed_forward_length` (OLMoE=1024,
        // Qwen2-MoE=5632 for the shared expert). `expert_feed_forward_length`
        // is optional. llama.cpp `qwen2moe.cpp` uses
        // `n_ff_exp = n_ff_exp ? n_ff_exp : n_ff / n_expert_used` (1408 for
        // Qwen1.5-MoE); the shared expert keeps the full `n_ff` (5632).
        let feed_forward_length = metadata_u64_any(file, &[key("feed_forward_length")]);
        let expert_ffn_dim = metadata_u64_any(file, &[key("expert_feed_forward_length")])
            .or_else(|| {
                feed_forward_length.map(|ff| {
                    if is_moe && n_experts_active > 0 {
                        ff / n_experts_active as u64
                    } else {
                        ff
                    }
                })
            })
            .unwrap_or_else(|| {
                best_effort_fields.push(
                    "moe.expert_ffn_dim (no expert_feed_forward_length/feed_forward_length; defaulted to 4x hidden_dim)",
                );
                (hidden_dim * 4) as u64
            }) as usize;
        let n_dense_leading_layers = if LEADING_DENSE_KEY_IS_INERT.contains(&arch.as_str()) {
            0
        } else {
            metadata_u64_any(file, &[key("leading_dense_block_count")]).unwrap_or(0) as usize
        };
        // The OTHER half of llama.cpp's dense-vs-MoE rule.
        // `ModelConfig::layer_is_dense` implements the leading-dense
        // prefix and not the `(il + 1) % n_moe_layer_step == 0` at
        // `src/models/ernie4-5-moe.cpp:64`, so a file whose step would
        // change the answer stops here rather than looking for expert
        // tensors on a layer that stores dense ones. `moe_interleave`
        // records what building the fixture found: llama.cpp cannot load
        // such a file either, because its own tensor loader has no step
        // in it.
        if let Some(reason) = crate::moe_interleave::interleave_step_refusal(
            &arch,
            metadata_u64_any(file, &[key("interleave_moe_layer_step")]),
        ) {
            return Err(LoadError::UnsupportedFeature(arch.clone(), reason));
        }

        // ik_llama.cpp's real gating-function hparam
        // (LLM_KV_EXPERT_GATING_FUNC: 1=softmax, 2=sigmoid) if the file
        // carries it; otherwise fall back to the same architecture-name
        // convention the hand-written presets in config.rs use (see
        // docs/MODELS.md for the citations behind that list).
        let gating = match metadata_u64_any(file, &[key("expert_gating_func")]) {
            Some(2) => GatingFunction::Sigmoid,
            Some(1) => GatingFunction::Softmax,
            _ => {
                if SIGMOID_GATING_ARCHITECTURES.contains(&arch.as_str()) {
                    GatingFunction::Sigmoid
                } else {
                    if is_moe {
                        best_effort_fields.push(
                            "moe.gating (no expert_gating_func key and architecture not in the known-sigmoid list; defaulted to softmax)",
                        );
                    }
                    GatingFunction::Softmax
                }
            }
        };

        // `{arch}.expert_weights_norm` (llama.cpp
        // `LLM_KV_EXPERT_WEIGHTS_NORM`) is the real metadata key for
        // whether the selected experts' weights are renormalised. Most
        // checkpoints do not carry it, which is why the fallback below
        // exists at all -- but when one does, the file's own answer wins
        // over an architecture-name guess.
        let norm_topk_prob = match file.metadata_bool(&key("expert_weights_norm")) {
            Some(v) => v,
            None => {
                // See `NO_TOPK_RENORMALIZE_ARCHITECTURES`'s doc comment:
                // an architecture-name lookup, the same convention
                // `gating`'s fallback above uses.
                if is_moe && matches!(gating, GatingFunction::Softmax) {
                    best_effort_fields.push(
                        "moe.norm_topk_prob (no expert_weights_norm key; defaulted by architecture-name lookup against NO_TOPK_RENORMALIZE_ARCHITECTURES)",
                    );
                }
                !NO_TOPK_RENORMALIZE_ARCHITECTURES.contains(&arch.as_str())
            }
        };

        // `{arch}.expert_weights_scale` (`LLM_KV_EXPERT_WEIGHTS_SCALE`).
        // llama.cpp's `build_moe_ffn` skips the multiply for both 0.0 and
        // 1.0, so both mean "no scaling" and both land on 1.0 here.
        let expert_weights_scale = metadata_f32_any(file, &[key("expert_weights_scale")])
            .filter(|s| *s != 0.0)
            .unwrap_or(1.0);

        // Real GGUF key (`{arch}.attention.sliding_window`, confirmed
        // against `gguf-py/gguf/constants.py`'s real
        // `LLM_KV_ATTENTION_SLIDING_WINDOW`). Some checkpoints
        // (confirmed for real published Qwen1.5-MoE/Qwen2-MoE GGUFs)
        // carry a nonzero window value even when the model's own
        // config disables sliding-window attention entirely
        // (`use_sliding_window: false`) -- llama.cpp's own convention
        // is that a window of 0 means "unused," so only a real nonzero
        // value here is treated as active.
        let sliding_window = metadata_u64_any(file, &[key("attention.sliding_window")])
            .map(|v| v as usize)
            .filter(|&w| w > 0)
            // `phi3` declares a window that llama.cpp deliberately does
            // NOT honour -- see `capability::swa_disabled_by_arch`. This
            // has to drop the window rather than pick a period, because
            // upstream is declining to use the file's value, not
            // choosing a different one.
            .filter(|_| !crate::capability::swa_disabled_by_arch(&arch));

        // Gemma alternating SWA period (`attention.sliding_window_pattern`).
        // llama.cpp: gemma2 defaults period=2, gemma3 defaults period=6 when
        // the pattern key is absent. A missing key must NOT mean "all SWA".
        //
        // The metadata key overrides the PERIOD only. The phase is a
        // property of the architecture in llama.cpp -- `dense_first` is
        // an argument to `set_swa_pattern`, not a GGUF key -- so it
        // comes from the registry either way.
        let swa_layout = crate::capability::default_swa_layout(&arch);
        let swa_dense_first = swa_layout.is_some_and(|p| p.dense_first);
        // llama.cpp reads this with `ml.get_key_or_arr`, so the value is
        // a scalar period OR an n_layer-long per-layer array. ferrox
        // carries one scalar `swa_pattern`, and `metadata_u64_any`
        // simply returns `None` for an array -- which silently
        // substituted `default_swa_layout`'s period for the layout the
        // file actually declared. Present-but-unreadable is the case to
        // refuse; presence alone is not, because the pattern itself is
        // implemented (`ModelConfig::layer_sliding_window`, both
        // phases). `capability::unsupported_feature_keys` used to refuse
        // presence alone, and its comment says why that was wrong.
        let swa_pattern_key = key("attention.sliding_window_pattern");
        if file.metadata(&swa_pattern_key).is_some()
            && metadata_u64_any(file, std::slice::from_ref(&swa_pattern_key)).is_none()
        {
            return Err(LoadError::UnsupportedFeature(
                arch.clone(),
                format!(
                    "{swa_pattern_key} is not a scalar period; llama.cpp accepts a \
                     per-layer array here (ml.get_key_or_arr) and ferrox carries one \
                     period for the whole model, so honouring it would mean substituting \
                     a different layout for the file's"
                ),
            ));
        }
        let swa_pattern = metadata_u64_any(file, &[swa_pattern_key])
            .map(|v| v as usize)
            .or_else(|| {
                sliding_window?;
                // llama.cpp hardcodes the period per architecture and
                // only lets the metadata key override it, so a missing
                // key is *not* "every layer windowed" -- see
                // `capability::default_swa_layout`.
                swa_layout.map(|p| p.period).or(
                    // Any Gemma variant not named in the table keeps the
                    // gemma3+ period rather than going uniform.
                    match arch_profile.family {
                        crate::capability::DecoderFamily::GemmaFamily => Some(6),
                        _ => None,
                    },
                )
            });

        let attn_logit_softcap = metadata_f32_any(
            file,
            &[
                key("attention.logit_softcapping"),
                key("attn_logit_softcapping"),
            ],
        )
        .filter(|&v| v > 0.0);
        let final_logit_softcap =
            metadata_f32_any(file, &[key("final_logit_softcapping")]).filter(|&v| v > 0.0);

        // Gemma: embeddings are scaled by sqrt(hidden_dim) at input.
        let embedding_scale = if matches!(
            arch_profile.family,
            crate::capability::DecoderFamily::GemmaFamily
        ) {
            Some((hidden_dim as f32).sqrt())
        } else {
            None
        };

        // llama.cpp's `f_attention_scale`, and ONLY where it differs from
        // the `1/sqrt(head_dim)` ferrox's attention kernels already
        // apply -- `Some` here means "pre-scale Q", so restating the
        // kernels' own scale would double-scale every score.
        //
        // For Gemma-2 and Gemma-3 that difference is real at 27B and
        // nowhere else (`capability::attention_scale_override` carries
        // the llama.cpp lines). This used to be a hardcoded `None` under
        // a comment that NAMED the 27B exception without implementing
        // it, so Gemma-2-27B scored 1.061x and Gemma-3-27B 1.146x too
        // large on every layer: a sharper softmax than the trained one,
        // fluent and wrong, with no error.
        let attention_scale = crate::capability::attention_scale_override(
            &arch, n_layers, hidden_dim, n_heads, head_dim,
        );

        // SWA-layer RoPE base. `llama_hparams` defaults it to 10000 and
        // the Gemma-3 lineage relies on that default; the architectures
        // in `swa_rope_base_follows_model` instead seed it from the
        // model's own base before the key can override.
        let rope_theta_swa = if sliding_window.is_some() {
            let fallback = if crate::capability::swa_rope_base_follows_model(&arch) {
                rope_theta
            } else {
                10_000.0
            };
            Some(
                metadata_f32_any(
                    file,
                    &[key("rope.freq_base_swa"), key("rope_freq_base_swa")],
                )
                .unwrap_or(fallback),
            )
        } else {
            None
        };

        let ffn_activation = match arch_profile.family {
            // Per-ARCHITECTURE first, because llama.cpp's choice is per
            // architecture and the family partition does not match it:
            // `grok` is StandardGqa and passes `LLM_FFN_GELU`.
            _ if crate::capability::uses_geglu(&arch) => crate::config::FfnActivation::Gelu,
            crate::capability::DecoderFamily::GemmaFamily => crate::config::FfnActivation::Gelu,
            crate::capability::DecoderFamily::PhiFamily => {
                crate::config::FfnActivation::SwigluFused
            }
            _ => crate::config::FfnActivation::Swiglu,
        };

        // Llama 3/3.1/3.2's real per-band RoPE frequency correction: one
        // model-level tensor (`TENSOR_NOT_REQUIRED`, `TENSOR_DUPLICATED`
        // for every layer but the first in the real llama.cpp source --
        // i.e. every layer shares this same array), not per-layer. See
        // `ferrox_core::attention::apply_rope_with_freq_factors`'s doc
        // comment for why this matters.
        let rope_freqs = load_f32_vec_optional(file, "rope_freqs.weight")?;

        // Phi-3/Phi-4 LongRoPE: two per-band factor tensors instead of
        // Llama's single `rope_freqs.weight`, selected by context size
        // (llama.cpp `llama_model::get_rope_factors`: `rope_freqs` wins if
        // present, else `rope_long` when the run's context exceeds
        // `rope.scaling.original_context_length`, else `rope_short`).
        //
        // Provisional pick from the checkpoint's advertised context length
        // (llama.cpp's default `n_ctx`). The definitive pick happens in
        // `ModelConfig::apply_runtime_context`, called from `ferrox run`
        // (`--ctx-size`) and from `verify_engine::load_and_tokenize`
        // (`n_tokens + 8`, matching `tools/llama_logits.c`).
        let rope_orig_ctx = metadata_u64_any(file, &[key("rope.scaling.original_context_length")])
            .map(|v| v as usize);
        // `rope_freqs.weight` outranks the LongRoPE pair (llama.cpp
        // `get_rope_factors` checks it first), so a checkpoint carrying
        // it never populates these and the runtime re-pick below cannot
        // overwrite a Llama-3 correction with a Phi one.
        let (rope_freqs_long, rope_freqs_short) = if rope_freqs.is_some() {
            (None, None)
        } else {
            (
                load_f32_vec_optional(file, "rope_factors_long.weight")?,
                load_f32_vec_optional(file, "rope_factors_short.weight")?,
            )
        };
        // Provisional pick from the checkpoint's own advertised context;
        // `ModelConfig::apply_runtime_context` re-picks once the run's
        // `--ctx-size` is known, which is the number llama.cpp decides on.
        let rope_freqs = match (rope_freqs, rope_orig_ctx) {
            (Some(f), _) => Some(f),
            (None, Some(orig)) => {
                let model_ctx = metadata_u64_any(file, &[key("context_length")])
                    .unwrap_or(orig as u64) as usize;
                if model_ctx > orig {
                    rope_freqs_long.clone().or_else(|| rope_freqs_short.clone())
                } else {
                    rope_freqs_short.clone().or_else(|| rope_freqs_long.clone())
                }
            }
            (None, None) => None,
        };

        // Partial rotary: only when the file says the rotary width is
        // narrower than a head. Equal values mean "whole head", which is
        // the same thing as `None` and stays `None` so nothing downstream
        // has to special-case it.
        let rope_dim = metadata_u64_any(file, &[key("rope.dimension_count")])
            .map(|d| d as usize)
            .filter(|d| *d > 0 && *d < head_dim);

        // See `ModelConfig::rope_attn_factor`.
        let rope_attn_factor = metadata_f32_any(file, &[key("rope.scaling.attn_factor")])
            .filter(|f| f.is_finite() && *f > 0.0)
            .unwrap_or(1.0);

        // YaRN long-context scaling. `rope.scaling.attn_factor` above is
        // only YaRN's *magnitude* term (ggml `rope_yarn`'s `mscale`); the
        // frequency half -- which bands get interpolated toward the
        // trained context and which stay extrapolated -- lives in
        // `rope.scaling.type` + `rope.scaling.factor`, and ferrox read
        // neither before this. A YaRN checkpoint was therefore roped as
        // if it declared no scaling at all: right near position 0 and
        // progressively wrong further in, i.e. the failure that reads as
        // long-prompt quality decay rather than as a bug.
        //
        // The rewrite is folded into `rope_freqs`, the same per-band
        // divisor array Llama-3's `rope_freqs.weight` supplies (ggml
        // divides each band's theta by it), so it rides the existing CPU
        // and Metal RoPE paths unchanged. When a file carries both, the
        // two corrections compose by multiplication, as they do in
        // llama.cpp (`ggml_rope_cache_init` divides by `freq_factors`
        // *and then* runs `rope_yarn`).
        // Linear scaling, which was silently DROPPED before this.
        //
        // `rope.scaling.type = "linear"` with factor s means rotating
        // position `p/s` instead of `p`. Since the angle is `p * freq`,
        // that is exactly `p * (freq / s)`, and `rope_freqs` already
        // divides each band's frequency. So a uniform vector of `s`
        // expresses it exactly and rides the existing CPU and Metal RoPE
        // paths unchanged, the same way YaRN does below.
        //
        // Before this, the type was compared against "yarn" and anything
        // else returned None, so a checkpoint declaring linear scaling
        // with factor 4 loaded and roped at UNSCALED positions where
        // llama.cpp divides them by 4. It answered as a different model
        // with no error. Affects the long-context community rescales
        // (`*-16k`, `*-32k` Llama-2 derivatives).

        // The file's own per-band factors, BEFORE any position-scaling
        // fold. That is what a sliding layer uses on an architecture
        // whose SWA layers do not inherit the trained scale -- llama.cpp
        // keeps the two apart as `freq_factors` (a tensor, the same for
        // every layer) and `freq_scale` (per layer,
        // `llama-model.cpp:2033`), while ferrox folds them into one
        // vector. See `config::RopeFreqs`.
        let rope_freqs_unscaled = rope_freqs.clone();

        let rope_freqs = match linear_scaling_from_gguf(file, &arch) {
            None => rope_freqs,
            Some(factor) => {
                let rotary_dim = rope_dim.unwrap_or(head_dim);
                if rotary_dim == 0 || !rotary_dim.is_multiple_of(2) {
                    best_effort_fields.push(
                        "rope_freqs (linear scaling declared but the rotary width is odd; \
                         scaling not applied)",
                    );
                    rope_freqs
                } else {
                    let linear = vec![factor; rotary_dim / 2];
                    match rope_freqs {
                        None => Some(linear),
                        // Compose by multiplication, as a file carrying
                        // its own `rope_freqs.weight` tensor and a
                        // declared linear factor means both.
                        Some(own) if own.len() == linear.len() => {
                            Some(own.iter().zip(linear.iter()).map(|(a, b)| a * b).collect())
                        }
                        Some(own) => {
                            best_effort_fields.push(
                                "rope_freqs (linear scaling declared but the file's own \
                                 rope_freqs tensor has a different width; scaling not applied)",
                            );
                            Some(own)
                        }
                    }
                }
            }
        };
        let rope_freqs = match yarn_scaling_from_gguf(file, &arch, rope_orig_ctx) {
            None => rope_freqs,
            Some(scaling) => {
                let rotary_dim = rope_dim.unwrap_or(head_dim);
                if rotary_dim == 0 || !rotary_dim.is_multiple_of(2) {
                    best_effort_fields.push(
                        "rope_freqs (YaRN declared but the rotary width is odd; scaling not applied)",
                    );
                    rope_freqs
                } else {
                    let yarn =
                        ferrox_core::attention::yarn_freq_factors(scaling, rotary_dim, rope_theta);
                    match rope_freqs {
                        None => Some(yarn),
                        Some(own) if own.len() == yarn.len() => {
                            Some(own.iter().zip(yarn.iter()).map(|(a, b)| a * b).collect())
                        }
                        Some(own) => {
                            best_effort_fields.push(
                                "rope_freqs (YaRN declared alongside a per-band factor tensor of a \
                                 different width; the file's own tensor is used unscaled)",
                            );
                            Some(own)
                        }
                    }
                }
            }
        };

        // The SWA half of the split. llama.cpp defaults
        // `rope_freq_scale_train_swa` to `1.0f`
        // (`src/llama-hparams.h:129`) and only the architectures in
        // `swa_rope_scale_follows_model` assign it from
        // `rope_freq_scale_train`; `get_rope_freq_scale`
        // (`llama-model.cpp:2033-2035`) then picks between them per
        // layer. `gemma3.cpp` is not on that list and its converter
        // writes the FULL-ATTENTION factor
        // (`conversion/base.py:1222-1230`), so a Gemma-3 4B/12B/27B was
        // rotating five layers in six at `p/8` where llama.cpp rotates
        // at `p`.
        //
        // "No scaling" is spelled as an all-ones divisor vector, which
        // is what dividing by nothing is, so the sliding layers need no
        // second code path anywhere downstream.
        let rope_freqs = rope_freqs.map(|full| {
            let swa = (sliding_window.is_some()
                && !crate::capability::swa_rope_scale_follows_model(&arch))
            .then(|| rope_freqs_unscaled.unwrap_or_else(|| vec![1.0; full.len()]))
            .filter(|swa| *swa != full);
            crate::config::RopeFreqs { full, swa }
        });

        // RoPE layout comes from the capability registry above (fail-
        // closed). Getting this wrong for `llama` (needs Norm) was the
        // real root cause of the Llama-3.1-8B early-stop/wrong-logits bug.

        if best_effort_fields.is_empty() {
            best_effort_fields.push(
                "none -- every field above was read directly from this file's own GGUF metadata",
            );
        }

        // LAST, deliberately. The generic path is a GUESS, so it has to
        // be opted into rather than fallen onto: it assumes plain GQA
        // with no ALiBi, no learned position embeddings and no
        // per-layer rope skipping, and that assumption was already
        // wrong for gpt2, mpt, refact, bloom and jais.
        //
        // But it runs AFTER every architecture-specific refusal, so a
        // checkpoint with a NAMED problem still reports that problem.
        // Checking first would have replaced "this uses ALiBi" with
        // "this is unaudited", which is true and much less useful.
        if matches!(
            arch_profile.path,
            crate::capability::ArchPath::GenericGqa { .. }
        ) && !crate::capability::is_audited_generic(&arch)
            && !matches!(
                std::env::var("FERROX_ALLOW_UNAUDITED_ARCH").ok().as_deref(),
                Some("1") | Some("true") | Some("on")
            )
        {
            return Err(LoadError::UnauditedArchitecture(
                arch.clone(),
                rope_layout,
                crate::capability::unaudited_refusal_detail(&arch),
            ));
        }

        Ok(ModelConfig {
            name,
            n_layers,
            hidden_dim,
            n_heads,
            n_kv_heads,
            head_dim,
            vocab_size,
            rope_theta,
            rms_norm_eps,
            // No GGUF file encodes a hybrid KDA/Gated-MLA attention
            // topology today; every real checkpoint loaded this way
            // runs the standard Gqa path.
            attention: crate::config::AttentionKind::Gqa,
            sliding_window,
            swa_pattern,
            swa_dense_first,
            moe: MoeLayerConfig {
                n_experts: n_experts.max(1),
                n_experts_active,
                n_shared_experts,
                hidden_dim,
                expert_ffn_dim,
                gating,
                norm_topk_prob,
                expert_group_count: metadata_u64_any(file, &[key("expert_group_count")])
                    .map(|v| v as usize)
                    .filter(|&c| c > 1),
                expert_group_used_count: metadata_u64_any(file, &[key("expert_group_used_count")])
                    .map(|v| v as usize)
                    .filter(|&c| c > 0),
                expert_weights_scale,
            },
            n_dense_leading_layers,
            rope_freqs,
            rope_layout,
            qk_norm_style,
            attn_logit_softcap,
            final_logit_softcap,
            embedding_scale,
            attention_scale,
            rope_attn_factor,
            rope_dim,
            rope_freqs_long,
            rope_freqs_short,
            rope_orig_ctx,
            rope_theta_swa,
            ffn_activation,
            best_effort_fields: Box::leak(best_effort_fields.into_boxed_slice()),
        })
    }
}

impl crate::sampling::RecommendedSampling {
    /// The sampling a GGUF recommends for itself, from the
    /// `general.sampling.*` metadata keys llama.cpp's converter writes
    /// when the source checkpoint carried a `generation_config.json`.
    ///
    /// This is the GGUF half of FreeToken's `load_generation_sampling`
    /// (`python/freetoken/utils/hf.py:92`), which checks the GGUF
    /// metadata *first* and only falls back to a `generation_config.json`
    /// sidecar for non-GGUF checkpoints -- a GGUF is a single file and
    /// has no sidecar to read.
    ///
    /// Key names are llama.cpp's own (`general.sampling.temp`, not
    /// `temperature`). Each key is independent: a file that names only
    /// `top_k` recommends only `top_k`, and the two fields it did not
    /// mention stay `None` so the server's own defaults keep speaking
    /// for them.
    ///
    /// `temp` / `top_p` are read as float *or* integer, because a
    /// converter that wrote `temp = 1` stores a GGUF integer and
    /// dropping that value would silently serve the checkpoint greedy --
    /// the exact repetition-loop failure the recommendation exists to
    /// prevent.
    pub fn from_gguf(file: &impl TensorSource) -> Self {
        let number = |k: &str| -> Option<f32> {
            file.metadata(k)
                .and_then(|v| v.as_f32().or_else(|| v.as_u64().map(|u| u as f32)))
        };
        crate::sampling::RecommendedSampling {
            temperature: number("general.sampling.temp"),
            top_p: number("general.sampling.top_p"),
            top_k: file
                .metadata("general.sampling.top_k")
                .and_then(|v| v.as_u64())
                .map(|v| v as usize),
        }
    }
}

/// The `linear` RoPE scaling factor, if this file declares one.
///
/// Deliberately separate from [`yarn_scaling_from_gguf`]: YaRN needs an
/// original context length and per-band betas, and linear needs neither.
/// Any factor at or below one is not a correction, and is treated as
/// absent rather than applied as a no-op.
fn linear_scaling_from_gguf(file: &impl TensorSource, arch: &str) -> Option<f32> {
    let key = |suffix: &str| format!("{arch}.{suffix}");
    let scaling_type = file.metadata_str(&key("rope.scaling.type"))?;
    if !scaling_type.eq_ignore_ascii_case("linear") {
        return None;
    }
    metadata_f32_any(file, &[key("rope.scaling.factor")]).filter(|f| f.is_finite() && *f > 1.0)
}

/// The YaRN RoPE scaling a GGUF declares, or `None` when this file
/// declares none that changes the rotation.
///
/// llama.cpp's key names (`llama-arch.cpp`
/// `LLM_KV_ROPE_SCALING_TYPE` / `_FACTOR`): `<arch>.rope.scaling.type`
/// is a string (`"none"`, `"linear"`, `"yarn"`, `"longrope"`) and
/// `<arch>.rope.scaling.factor` the ratio of served to trained context.
/// `beta_fast` / `beta_slow` are read from both the plain and the
/// `yarn_`-prefixed spelling and otherwise fall back to the reference's
/// own defaults (32.0 / 1.0), which is what a real checkpoint relies on
/// -- almost none of them write those two keys.
///
/// `None` is returned for every case where applying YaRN would be a
/// guess or a no-op rather than a correction, so that no checkpoint's
/// rotation moves without the file having asked for it:
///
/// * a scaling type other than `yarn` (`linear` divides positions,
///   `longrope` rides the `rope_factors_long`/`_short` tensors this
///   loader already reads -- neither is this rewrite, and treating them
///   as YaRN would rope them wrong in a *new* way instead of leaving
///   them as they are),
/// * a missing, non-finite or `<= 1.0` factor (the reference's own
///   `get_mscale` treats `scale <= 1` as unscaled, and a factor of 1.0
///   makes every band's divisor exactly 1.0 anyway),
/// * a missing `rope.scaling.original_context_length` -- the trained
///   context is what the correction range is measured against, and
///   inventing one (say, from `context_length`, which on a YaRN file is
///   the *extended* length) would put the ramp in the wrong place and
///   quietly rope the checkpoint at frequencies nobody trained.
fn yarn_scaling_from_gguf(
    file: &impl TensorSource,
    arch: &str,
    orig_ctx: Option<usize>,
) -> Option<ferrox_core::attention::YarnScaling> {
    let key = |suffix: &str| format!("{arch}.{suffix}");
    let scaling_type = file.metadata_str(&key("rope.scaling.type"))?;
    if !scaling_type.eq_ignore_ascii_case("yarn") {
        return None;
    }
    let factor = metadata_f32_any(file, &[key("rope.scaling.factor")])
        .filter(|f| f.is_finite() && *f > 1.0)?;
    let orig_max_pos = orig_ctx?;
    let beta = |suffix: &str, default: f32| -> f32 {
        metadata_f32_any(
            file,
            &[
                key(&format!("rope.scaling.{suffix}")),
                key(&format!("rope.scaling.yarn_{suffix}")),
            ],
        )
        .filter(|v| v.is_finite() && *v > 0.0)
        .unwrap_or(default)
    };
    Some(ferrox_core::attention::YarnScaling {
        factor,
        beta_fast: beta("beta_fast", 32.0),
        beta_slow: beta("beta_slow", 1.0),
        orig_max_pos,
        // No GGUF key carries the reference's `truncate` flag, and its
        // default is `true`; a file that wanted the fractional range
        // would have no way to say so here.
        truncate: true,
    })
}

pub(crate) fn find_info<'a>(
    file: &'a impl TensorSource,
    name: &str,
) -> Result<&'a TensorInfo, LoadError> {
    file.find_tensor(name)
        .ok_or_else(|| LoadError::Gguf(GgufError::TensorNotFound(name.to_string())))
}

/// Like `load_f32_vec`, but for tensors that only exist on some
/// checkpoints (e.g. `attn_q_norm`/`attn_k_norm` -- OLMoE-style
/// per-projection QK-RMSNorm applied to the full q_proj/k_proj output
/// before RoPE, confirmed against `OlmoeAttention.forward` in
/// `transformers/models/olmoe/modeling_olmoe.py`: `q_norm(q_proj(x))`,
/// `k_norm(k_proj(x))`, both plain RMSNorm over the whole projected
/// width, not per-head). Absent for every other preset/fixture this
/// loader already handles -- `None` there is correct, not a missing
/// feature.
/// Loads the five gpt-oss-only tensors for one layer.
///
/// Every one of them is **required**: a gpt-oss checkpoint that is
/// missing any of these is not a gpt-oss checkpoint ferrox can run, and
/// quietly substituting zeros would reintroduce exactly the
/// silently-wrong-graph failure this path exists to remove. The lengths
/// are asserted against the config for the same reason -- a bias of the
/// wrong width would otherwise be applied to a `zip`-truncated prefix
/// and produce a plausible, wrong answer.
///
/// Shapes follow `src/models/openai-moe.cpp::load_arch_tensors`:
/// `attn_sinks {n_head}`, `attn_output.bias {n_embd}`,
/// `ffn_gate_inp.bias {n_expert}`, `ffn_{gate,up}_exps.bias
/// {n_ff_exp, n_expert}`, `ffn_down_exps.bias {n_embd, n_expert}`.
/// GGUF stores the fastest dimension first, so the 2-D bias tensors
/// arrive expert-major and split by simple chunking.
fn load_gpt_oss_layer(
    file: &impl TensorSource,
    l: usize,
    config: &ModelConfig,
) -> Result<crate::decoder::GptOssLayer, LoadError> {
    let n_experts = config.moe.n_experts;
    let ff = config.moe.expert_ffn_dim;

    let want = |name: &str, got: usize, expect: usize| -> Result<(), LoadError> {
        if got == expect {
            Ok(())
        } else {
            Err(LoadError::UnsupportedFeature(
                config.name.to_string(),
                format!("{name} has {got} elements, expected {expect}"),
            ))
        }
    };

    let attn_sinks = load_f32_vec(file, &format!("blk.{l}.attn_sinks.weight"))?;
    want(
        &format!("blk.{l}.attn_sinks.weight"),
        attn_sinks.len(),
        config.n_heads,
    )?;
    let o_bias = load_f32_vec(file, &format!("blk.{l}.attn_output.bias"))?;
    want(
        &format!("blk.{l}.attn_output.bias"),
        o_bias.len(),
        config.hidden_dim,
    )?;
    let router_bias = load_f32_vec(file, &format!("blk.{l}.ffn_gate_inp.bias"))?;
    want(
        &format!("blk.{l}.ffn_gate_inp.bias"),
        router_bias.len(),
        n_experts,
    )?;

    let gate_b = load_f32_vec(file, &format!("blk.{l}.ffn_gate_exps.bias"))?;
    want(
        &format!("blk.{l}.ffn_gate_exps.bias"),
        gate_b.len(),
        n_experts * ff,
    )?;
    let up_b = load_f32_vec(file, &format!("blk.{l}.ffn_up_exps.bias"))?;
    want(
        &format!("blk.{l}.ffn_up_exps.bias"),
        up_b.len(),
        n_experts * ff,
    )?;
    let down_b = load_f32_vec(file, &format!("blk.{l}.ffn_down_exps.bias"))?;
    want(
        &format!("blk.{l}.ffn_down_exps.bias"),
        down_b.len(),
        n_experts * config.hidden_dim,
    )?;

    let expert_bias = (0..n_experts)
        .map(|e| ferrox_moe::ExpertBias {
            gate: gate_b[e * ff..(e + 1) * ff].to_vec(),
            up: up_b[e * ff..(e + 1) * ff].to_vec(),
            down: down_b[e * config.hidden_dim..(e + 1) * config.hidden_dim].to_vec(),
        })
        .collect();

    Ok(crate::decoder::GptOssLayer {
        attn_sinks,
        o_bias,
        router_bias,
        expert_bias,
    })
}

pub(crate) fn load_f32_vec_optional(
    file: &impl TensorSource,
    name: &str,
) -> Result<Option<Vec<f32>>, LoadError> {
    if file.find_tensor(name).is_none() {
        return Ok(None);
    }
    Ok(Some(load_f32_vec(file, name)?))
}

/// A norm weight that some converters spell `<base>.weight` and others
/// spell just `<base>`.
///
/// This exists for exactly one pair of tensors, `post_attention_norm`
/// and `post_ffw_norm`, and it is this repo's dominant bug shape in
/// llama.cpp's own trees: TWO SPELLINGS OF ONE NAME, WITH NOTHING
/// ENFORCING AGREEMENT.
///
/// `LLM_TN` appends `.weight` only when it is given a suffix
/// (`src/llama-arch.cpp:898-910`). Every architecture that creates these
/// two tensors passes one -- `tn(LLM_TENSOR_ATTN_POST_NORM, "weight",
/// i)` in gemma2, gemma3, glm4, exaone4, afmoe and the rest -- EXCEPT
/// `plamo3`, which uses the two-argument overload
/// (`src/models/plamo3.cpp:52,55`) and therefore asks for
/// `blk.N.post_attention_norm` with no suffix at all.
///
/// The converter agrees with it, by a second accident that happens to
/// line up: `gguf-py/gguf/tensor_mapping.py:368,434` give the PLaMo
/// entries as `model.layers.layers.{bid}.post_mixer_norm.weight` and
/// `...post_mlp_norm.weight` -- keys that already END in `.weight`.
/// `TensorNameMap.get_type_and_name` (:2585-2594) tries an exact match
/// FIRST and only falls back to stripping a suffix, so those two match
/// exactly and the mapped name is emitted with nothing appended.
///
/// So a real PLaMo-3 GGUF carries `blk.N.post_attention_norm` and
/// `blk.N.post_ffw_norm`, and every Gemma-lineage GGUF carries the same
/// two names with `.weight`. Reading only one spelling means one of the
/// two families always fails on a missing tensor. ferrox read only
/// `.weight`, which is why `plamo3` could not have loaded a real
/// checkpoint -- fail-closed rather than wrong, but not "a fixture
/// away", which is what its triage verdict said.
///
/// Both spellings are accepted rather than one being chosen per
/// architecture, because the choice is a property of the file and
/// nothing in the metadata declares it. Neither present is still
/// `None`.
pub(crate) fn load_norm_vec_either_spelling(
    file: &impl TensorSource,
    base: &str,
) -> Result<Option<Vec<f32>>, LoadError> {
    let suffixed = format!("{base}.weight");
    if file.find_tensor(&suffixed).is_some() {
        return load_f32_vec_optional(file, &suffixed);
    }
    load_f32_vec_optional(file, base)
}

/// Slice `n` rows starting at `start` out of a quantized matrix without
/// dequantizing: every `Quantized` kind stores one interleaved block
/// buffer per row (fixed `row_bytes`), so a row range is a contiguous
/// byte range. Mapped sources stay zero-copy (sub-range of the same
/// mmap); other backings get an owned copy. Returns `None` for non-
/// quantized matrices (F32 / MXFP4) -- callers fall back to dequant.
fn slice_quantized_rows(m: &WeightMatrix, start: usize, n: usize) -> Option<WeightMatrix> {
    let WeightMatrix::Quantized {
        data,
        rows,
        cols,
        kind,
    } = m
    else {
        return None;
    };
    let total = data.len();
    if *rows == 0 || total % *rows != 0 || start + n > *rows {
        return None;
    }
    let row_bytes = total / *rows;
    let (b0, b1) = (start * row_bytes, (start + n) * row_bytes);
    let bytes = match data {
        WeightBytes::Mapped { mmap, range } => WeightBytes::Mapped {
            mmap: mmap.clone(),
            range: range.start + b0..range.start + b1,
        },
        other => WeightBytes::Owned(other.as_slice()[b0..b1].to_vec()),
    };
    Some(WeightMatrix::Quantized {
        data: bytes,
        rows: n,
        cols: *cols,
        kind: *kind,
    })
}

/// Loads Q/K/V projections: prefers split `attn_{q,k,v}.weight`, falls
/// back to fused `attn_qkv.weight` (Phi-3 / some Qwen GGUFs) by
/// slicing quantized rows (zero-copy for mmapped GGUFs; dequant only
/// for non-quantized storage). Mirrors llama.cpp `create_tensor_qkv`.
fn load_qkv_projections(
    file: &impl TensorSource,
    layer: usize,
    config: &ModelConfig,
) -> Result<(WeightMatrix, WeightMatrix, WeightMatrix), LoadError> {
    let q_name = format!("blk.{layer}.attn_q.weight");
    let k_name = format!("blk.{layer}.attn_k.weight");
    let v_name = format!("blk.{layer}.attn_v.weight");
    let fused_name = format!("blk.{layer}.attn_qkv.weight");

    if file.find_tensor(&q_name).is_some() {
        return Ok((
            load_weight_matrix(file, &q_name)?,
            load_weight_matrix(file, &k_name)?,
            load_weight_matrix(file, &v_name)?,
        ));
    }
    if file.find_tensor(&fused_name).is_none() {
        return Err(LoadError::Gguf(GgufError::TensorNotFound(q_name)));
    }

    let fused = load_weight_matrix(file, &fused_name)?;
    let q_rows = config.n_heads * config.head_dim;
    let kv_rows = config.n_kv_heads * config.head_dim;
    let expected = q_rows + 2 * kv_rows;
    if fused.rows() != expected {
        // Phi-3 sometimes stores Q as full n_embd (== q_rows when MHA).
        return Err(LoadError::UnsupportedFeature(
            config.name.to_string(),
            format!(
                "{fused_name} has {} rows; expected q+k+v = {} \
                 (n_heads*head_dim + 2*n_kv_heads*head_dim)",
                fused.rows(),
                expected
            ),
        ));
    }
    let cols = fused.cols();
    // Quantized fused tensor: split by row ranges without dequantizing,
    // keeping Q/K/V on the quantized (Metal-capable) matvec path.
    if let (Some(q), Some(k), Some(v)) = (
        slice_quantized_rows(&fused, 0, q_rows),
        slice_quantized_rows(&fused, q_rows, kv_rows),
        slice_quantized_rows(&fused, q_rows + kv_rows, kv_rows),
    ) {
        return Ok((q, k, v));
    }
    // Non-quantized storage: dequant once and split.
    let mut full = Vec::with_capacity(fused.rows() * cols);
    for r in 0..fused.rows() {
        full.extend_from_slice(&fused.dequant_row(r));
    }
    let q = WeightMatrix::F32(Tensor::new(
        full[..q_rows * cols].to_vec(),
        vec![q_rows, cols],
    ));
    let k = WeightMatrix::F32(Tensor::new(
        full[q_rows * cols..(q_rows + kv_rows) * cols].to_vec(),
        vec![kv_rows, cols],
    ));
    let v = WeightMatrix::F32(Tensor::new(
        full[(q_rows + kv_rows) * cols..].to_vec(),
        vec![kv_rows, cols],
    ));
    Ok((q, k, v))
}

/// Dense-layer FFN tensors: standard gate/up/down, or Phi-3 fused
/// `ffn_up` with `2 * expert_ffn_dim` rows and no separate gate.
fn load_dense_expert(
    file: &impl TensorSource,
    layer: usize,
    config: &ModelConfig,
) -> Result<ExpertWeights, LoadError> {
    let gate_name = format!("blk.{layer}.ffn_gate.weight");
    let up_name = format!("blk.{layer}.ffn_up.weight");
    let down_name = format!("blk.{layer}.ffn_down.weight");
    if file.find_tensor(&gate_name).is_some() {
        return Ok(ExpertWeights {
            gate: load_weight_matrix(file, &gate_name)?,
            up: load_weight_matrix(file, &up_name)?,
            down: load_weight_matrix(file, &down_name)?,
        });
    }
    // Phi-3 fused SwiGLU: up is [hidden, 2*ff], first half gate, second up.
    let fused = load_weight_matrix(file, &up_name)?;
    let ff = config.moe.expert_ffn_dim;
    if fused.rows() != 2 * ff {
        return Err(LoadError::UnsupportedFeature(
            config.name.to_string(),
            format!(
                "{up_name} has {} rows without a companion ffn_gate; \
                 expected fused SwiGLU with 2*ffn_dim = {} rows",
                fused.rows(),
                2 * ff
            ),
        ));
    }
    let cols = fused.cols();
    // Quantized fused gate+up: split by rows, no dequant (Metal-capable).
    if let (Some(gate), Some(up)) = (
        slice_quantized_rows(&fused, 0, ff),
        slice_quantized_rows(&fused, ff, ff),
    ) {
        return Ok(ExpertWeights {
            gate,
            up,
            down: load_weight_matrix(file, &down_name)?,
        });
    }
    let mut full = Vec::with_capacity(fused.rows() * cols);
    for r in 0..fused.rows() {
        full.extend_from_slice(&fused.dequant_row(r));
    }
    let gate = WeightMatrix::F32(Tensor::new(full[..ff * cols].to_vec(), vec![ff, cols]));
    let up = WeightMatrix::F32(Tensor::new(full[ff * cols..].to_vec(), vec![ff, cols]));
    Ok(ExpertWeights {
        gate,
        up,
        down: load_weight_matrix(file, &down_name)?,
    })
}

/// Widen a raw plain-float tensor (`F32` / `F16` / `BF16`) to `f32`.
///
/// The three unquantized element types are handled identically at every
/// call site (eager widening to an owned buffer -- none of them has a
/// block structure a fused dot kernel could exploit), and each of the
/// seven GGUF loaders used to inline the same two-way match. F16 had no
/// arm in any of them, which made every `*-f16.gguf` a hard
/// `UnsupportedDtype` even though the type was parsed and sized.
pub(crate) fn widen_plain_float(
    dtype: GgmlType,
    raw: &[u8],
    name: &str,
) -> Result<Vec<f32>, LoadError> {
    match dtype {
        GgmlType::F32 => {
            let mut out = Vec::with_capacity(raw.len() / 4);
            for chunk in raw.as_chunks::<4>().0 {
                out.push(f32::from_le_bytes(*chunk));
            }
            Ok(out)
        }
        GgmlType::F16 => ferrox_quant::dequant_f16(raw)
            .map_err(|_| LoadError::UnsupportedDtype(name.to_string(), GgmlType::F16)),
        GgmlType::BF16 => ferrox_quant::dequant_bf16(raw)
            .map_err(|_| LoadError::UnsupportedDtype(name.to_string(), GgmlType::BF16)),
        // MXFP4 is accepted as a weight matrix and as an MoE expert
        // tensor, and `WeightMatrix::dequant` already calls this
        // dequantizer, so refusing it here made a 1-D MXFP4 norm or
        // bias a hard load error on a checkpoint whose 2-D tensors of
        // the same type load fine. That contradicted this function's
        // own contract, which is to widen whatever the loaders accept.
        GgmlType::MXFP4 => ferrox_quant::dequant_mxfp4_gguf(raw)
            .map_err(|_| LoadError::UnsupportedDtype(name.to_string(), GgmlType::MXFP4)),
        other => Err(LoadError::UnsupportedDtype(name.to_string(), other)),
    }
}

pub(crate) fn load_f32_vec(file: &impl TensorSource, name: &str) -> Result<Vec<f32>, LoadError> {
    let info = find_info(file, name)?;
    let raw = file.tensor_bytes(name)?;
    match info.dtype {
        // MXFP4 rides with the plain floats because `widen_plain_float`
        // is where its arm already lives -- routing it here rather than
        // giving this table its own `dequant_mxfp4_gguf` call keeps ONE
        // MXFP4 arm in this file instead of two that can drift.
        //
        // It has to be in *both* tables' reach, and it was in neither's:
        // `load_weight_matrix` accepts MXFP4 as a 2-D weight and
        // `load_moe_expert_matrices` accepts it as an expert tensor, so
        // a checkpoint whose norms happen to be MXFP4 failed here with
        // `UnsupportedDtype` while its far larger tensors of the exact
        // same dtype loaded fine.
        GgmlType::F32 | GgmlType::F16 | GgmlType::BF16 | GgmlType::MXFP4 => {
            widen_plain_float(info.dtype, raw, name)
        }
        GgmlType::Q8_0 => ferrox_quant::dequant_q8_0(raw)
            .map_err(|_| LoadError::UnsupportedDtype(name.to_string(), GgmlType::Q8_0)),
        GgmlType::Q4_0 => ferrox_quant::dequant_q4_0(raw)
            .map_err(|_| LoadError::UnsupportedDtype(name.to_string(), GgmlType::Q4_0)),
        GgmlType::Q4K => ferrox_quant::dequant_q4_k(raw)
            .map_err(|_| LoadError::UnsupportedDtype(name.to_string(), GgmlType::Q4K)),
        GgmlType::Q5K => ferrox_quant::dequant_q5_k(raw)
            .map_err(|_| LoadError::UnsupportedDtype(name.to_string(), GgmlType::Q5K)),
        GgmlType::Q6K => ferrox_quant::dequant_q6_k(raw)
            .map_err(|_| LoadError::UnsupportedDtype(name.to_string(), GgmlType::Q6K)),
        GgmlType::Q2K => ferrox_quant::dequant_q2_k(raw)
            .map_err(|_| LoadError::UnsupportedDtype(name.to_string(), GgmlType::Q2K)),
        GgmlType::Q3K => ferrox_quant::dequant_q3_k(raw)
            .map_err(|_| LoadError::UnsupportedDtype(name.to_string(), GgmlType::Q3K)),
        GgmlType::Q4_1 => ferrox_quant::dequant_q4_1(raw)
            .map_err(|_| LoadError::UnsupportedDtype(name.to_string(), GgmlType::Q4_1)),
        GgmlType::Q5_0 => ferrox_quant::dequant_q5_0(raw)
            .map_err(|_| LoadError::UnsupportedDtype(name.to_string(), GgmlType::Q5_0)),
        GgmlType::Q5_1 => ferrox_quant::dequant_q5_1(raw)
            .map_err(|_| LoadError::UnsupportedDtype(name.to_string(), GgmlType::Q5_1)),
        GgmlType::Q8_1 => ferrox_quant::dequant_q8_1(raw)
            .map_err(|_| LoadError::UnsupportedDtype(name.to_string(), GgmlType::Q8_1)),
        GgmlType::IQ4NL => ferrox_quant::dequant_iq4_nl(raw)
            .map_err(|_| LoadError::UnsupportedDtype(name.to_string(), GgmlType::IQ4NL)),
        GgmlType::IQ4XS => ferrox_quant::dequant_iq4_xs(raw)
            .map_err(|_| LoadError::UnsupportedDtype(name.to_string(), GgmlType::IQ4XS)),
        // The codebook-grid tiers. Rare on the 1-D tensors this
        // function widens (norms and biases are almost always F32),
        // but a dtype ferrox can decode should never be rejected here
        // just because the *other* dispatch table below knows it --
        // that split is how a supported format turns into a load
        // failure on the one checkpoint that uses it.
        GgmlType::IQ1S => ferrox_quant::dequant_iq1_s(raw)
            .map_err(|_| LoadError::UnsupportedDtype(name.to_string(), GgmlType::IQ1S)),
        GgmlType::IQ1M => ferrox_quant::dequant_iq1_m(raw)
            .map_err(|_| LoadError::UnsupportedDtype(name.to_string(), GgmlType::IQ1M)),
        GgmlType::IQ2XXS => ferrox_quant::dequant_iq2_xxs(raw)
            .map_err(|_| LoadError::UnsupportedDtype(name.to_string(), GgmlType::IQ2XXS)),
        GgmlType::IQ2XS => ferrox_quant::dequant_iq2_xs(raw)
            .map_err(|_| LoadError::UnsupportedDtype(name.to_string(), GgmlType::IQ2XS)),
        GgmlType::IQ2S => ferrox_quant::dequant_iq2_s(raw)
            .map_err(|_| LoadError::UnsupportedDtype(name.to_string(), GgmlType::IQ2S)),
        GgmlType::IQ3XXS => ferrox_quant::dequant_iq3_xxs(raw)
            .map_err(|_| LoadError::UnsupportedDtype(name.to_string(), GgmlType::IQ3XXS)),
        GgmlType::IQ3S => ferrox_quant::dequant_iq3_s(raw)
            .map_err(|_| LoadError::UnsupportedDtype(name.to_string(), GgmlType::IQ3S)),
        other => Err(LoadError::UnsupportedDtype(name.to_string(), other)),
    }
}

/// Loads a 2D weight matrix, keeping Q8_0/Q4_0 tensors quantized (raw
/// bytes copied out, never dequantized) and only expanding truly F32
/// tensors. This is the memory- and bandwidth-saving path: for a
/// multi-billion-parameter checkpoint the difference between this and
/// "dequant everything on load" is the difference between fitting in
/// RAM and not.
pub(crate) fn load_weight_matrix(
    file: &impl TensorSource,
    name: &str,
) -> Result<WeightMatrix, LoadError> {
    let info = find_info(file, name)?;
    // GGUF's on-disk `ne[]` shape array is fastest-varying-dimension-first
    // (ggml convention), i.e. `[in_features, out_features]` for a 2D
    // weight matrix -- the *reverse* of the row-major `[rows, cols]` =
    // `[out_features, in_features]` order `WeightMatrix`/`matmul_f32`
    // need. Reversed here once so every consumer below gets the correct
    // orientation. Before this reversal existed, every 2D tensor in an
    // externally-produced GGUF file was silently loaded transposed -- a
    // real bug found by running a real downloaded checkpoint
    // (TinyLlama-1.1B-Chat, e.g. `attn_k.weight`'s real raw shape is
    // `[2048, 256]` = `[hidden_dim, kv_dim]` = `[in, out]`) -- found
    // as a real transposition bug affecting every externally-produced
    // GGUF file, caught by serving a real downloaded checkpoint.
    let shape: Vec<usize> = info.shape.iter().rev().map(|&d| d as usize).collect();
    // A ggml tensor's `ne[]` is always four long and trailing 1s are
    // implicit, so a GGUF writer is free to store a `[in, 1]` matrix
    // with `n_dims = 1`. llama.cpp reads it back as a matrix anyway --
    // `check_tensor_dims` compares each requested dimension against
    // `cur->ne[i]` and requires 1 for the dimensions the file does not
    // carry -- so a single-output projection is a 2-D weight there and
    // must be one here. Refusing it instead made a real checkpoint
    // unloadable: `cross-encoder/ms-marco-MiniLM-L6-v2` writes
    // `cls.output.weight` as `[384]`, i.e. the 1x384 relevance head
    // that `/v1/rerank` exists to run, and the whole route died at load
    // with "expected 2D".
    let (rows, cols) = match shape.as_slice() {
        [r, c] => (*r, *c),
        [c] => (1, *c),
        other => {
            return Err(LoadError::UnsupportedDtype(
                format!("{name} (expected 2D, got shape {other:?})"),
                info.dtype,
            ))
        }
    };
    // `shape` is what the file said; `[rows, cols]` is what the matrix
    // is. They differ exactly in the 1-D case above, and the `Tensor`
    // must carry the matrix shape or `apply` reads it as a vector.
    let shape = vec![rows, cols];

    match info.dtype {
        // BF16 has no block/scale structure to keep quantized-in-place
        // the way Q4_0/Q8_0/K-quants do -- there's no fused dot kernel
        // that would make sense for a plain narrowed float, so it's
        // eagerly widened to an owned f32 Tensor exactly like F32
        // tensors already are.
        GgmlType::F32 | GgmlType::F16 | GgmlType::BF16 => {
            let data = load_f32_vec(file, name)?;
            Ok(WeightMatrix::F32(Tensor::new(data, shape)))
        }
        other => match quant_kind_for(other) {
            Some(kind) => {
                let (mmap, range) = file.tensor_mapped_range(name)?;
                #[cfg(feature = "metal")]
                ferrox_metal::gpu::register_weight_mmap(Arc::clone(&mmap));
                Ok(WeightMatrix::Quantized {
                    data: WeightBytes::Mapped { mmap, range },
                    rows,
                    cols,
                    kind,
                })
            }
            None => Err(LoadError::UnsupportedDtype(name.to_string(), other)),
        },
    }
}

/// Splits a packed 3D MoE expert tensor `blk.N.ffn_{gate,up,down}_exps.weight`
/// (shape `[n_experts, out_dim, in_dim]`) into per-expert `WeightMatrix`es,
/// slicing raw bytes directly (quantized tensors stay quantized; block
/// boundaries never cross expert boundaries since `in_dim` is a whole
/// number of quantization blocks). Matches llama.cpp/ik_llama.cpp layout
/// confirmed on real OLMoE and Qwen2-MoE GGUF checkpoints.
pub(crate) fn split_expert_tensor(
    file: &impl TensorSource,
    name: &str,
    n_experts: usize,
) -> Result<Vec<WeightMatrix>, LoadError> {
    let info = find_info(file, name)?;
    // Real raw shape is `[in_dim, out_dim, n_experts]` (ggml's
    // fastest-first `ne[]` order -- see `load_weight_matrix`'s doc
    // comment for the confirmed 2D case this generalizes from). `n_experts`
    // is the slowest-varying (last, i.e. outermost/most-major) dimension,
    // so each expert's `out_dim*in_dim` block is contiguous with experts
    // back-to-back in the mmap.
    if info.shape.len() != 3 || info.shape[2] as usize != n_experts {
        let file_experts = info.shape.last().map(|&d| d as usize).unwrap_or(0);
        return Err(LoadError::ExpertCountMismatch(
            name.to_string(),
            file_experts,
            n_experts,
        ));
    }
    let out_dim = info.shape[1] as usize;
    let in_dim = info.shape[0] as usize;
    let raw = file.tensor_bytes(name)?;

    match info.dtype {
        GgmlType::F32 | GgmlType::F16 | GgmlType::BF16 => {
            let all = crate::loader::widen_plain_float(info.dtype, raw, name)?;
            let per_expert = out_dim * in_dim;
            Ok((0..n_experts)
                .map(|e| {
                    WeightMatrix::F32(Tensor::new(
                        all[e * per_expert..(e + 1) * per_expert].to_vec(),
                        vec![out_dim, in_dim],
                    ))
                })
                .collect())
        }
        other => match quant_kind_for(other) {
            Some(kind) => {
                let (mmap, full_range) = file.tensor_mapped_range(name)?;
                #[cfg(feature = "metal")]
                ferrox_metal::gpu::register_weight_mmap(Arc::clone(&mmap));
                let bytes_per_expert = raw.len() / n_experts;
                Ok((0..n_experts)
                    .map(|e| WeightMatrix::Quantized {
                        data: WeightBytes::Mapped {
                            mmap: Arc::clone(&mmap),
                            range: (full_range.start + e * bytes_per_expert)
                                ..(full_range.start + (e + 1) * bytes_per_expert),
                        },
                        rows: out_dim,
                        cols: in_dim,
                        kind,
                    })
                    .collect())
            }
            None => Err(LoadError::UnsupportedDtype(name.to_string(), other)),
        },
    }
}

/// When every routed expert is mmap-backed with a Metal simdgroup-GEMM
/// kind and back-to-back slices, record the combined gate/up/down planes
/// for Metal packed MoE. Gate/up/down may differ in kind (e.g. Q4_K /
/// Q4_K / Q8_0) but must be uniform across experts per role.
#[cfg(feature = "metal")]
fn try_build_moe_packed_q4_planes(experts: &[ExpertWeights]) -> Option<MoePackedQ4Planes> {
    use ferrox_core::weight_matrix::{QuantKind, WeightBytes};
    use std::sync::Arc;

    if experts.is_empty() {
        return None;
    }

    fn mapped_sg(m: &WeightMatrix) -> Option<(WeightBytes, usize, &'static str)> {
        match m {
            WeightMatrix::Quantized {
                data: WeightBytes::Mapped { mmap, range },
                rows,
                kind,
                ..
            } => {
                let kind_str = match kind {
                    QuantKind::Q4_0 => "Q4_0",
                    QuantKind::Q5_0 => "Q5_0",
                    QuantKind::Q4K => "Q4_K",
                    QuantKind::Q5K => "Q5_K",
                    QuantKind::Q6K => "Q6_K",
                    QuantKind::Q8_0 => "Q8_0",
                    QuantKind::IQ4XS => "IQ4_XS",
                    _ => return None,
                };
                let _ = ferrox_metal::gpu::mul_mm_sg_meta(kind_str)?;
                Some((
                    WeightBytes::Mapped {
                        mmap: Arc::clone(mmap),
                        range: range.clone(),
                    },
                    *rows,
                    kind_str,
                ))
            }
            _ => None,
        }
    }

    let (gate0, ffn_rows, gate_kind) = mapped_sg(&experts[0].gate)?;
    let (up0, up_rows, up_kind) = mapped_sg(&experts[0].up)?;
    let (down0, hidden_rows, down_kind) = mapped_sg(&experts[0].down)?;
    if up_rows != ffn_rows {
        return None;
    }
    let WeightBytes::Mapped {
        mmap: gate_mmap,
        range: gate0_range,
    } = &gate0
    else {
        return None;
    };
    let WeightBytes::Mapped {
        mmap: up_mmap,
        range: up0_range,
    } = &up0
    else {
        return None;
    };
    let WeightBytes::Mapped {
        mmap: down_mmap,
        range: down0_range,
    } = &down0
    else {
        return None;
    };

    let gate_stride = gate0_range.len();
    let up_stride = up0_range.len();
    let down_stride = down0_range.len();
    if gate_stride == 0 || up_stride == 0 || down_stride == 0 {
        return None;
    }

    let n = experts.len();
    for (i, ex) in experts.iter().enumerate().skip(1) {
        let (g, fr, gk) = mapped_sg(&ex.gate)?;
        let (u, ur, uk) = mapped_sg(&ex.up)?;
        let (d, hr, dk) = mapped_sg(&ex.down)?;
        if gk != gate_kind || uk != up_kind || dk != down_kind {
            return None;
        }
        let WeightBytes::Mapped { mmap, range } = &g else {
            return None;
        };
        if fr != ffn_rows {
            return None;
        }
        if !Arc::ptr_eq(mmap, gate_mmap)
            || range.len() != gate_stride
            || range.start != gate0_range.start + i * gate_stride
        {
            return None;
        }
        let WeightBytes::Mapped { mmap, range } = &u else {
            return None;
        };
        if ur != ffn_rows
            || !Arc::ptr_eq(mmap, up_mmap)
            || range.len() != up_stride
            || range.start != up0_range.start + i * up_stride
        {
            return None;
        }
        let WeightBytes::Mapped { mmap, range } = &d else {
            return None;
        };
        if hr != hidden_rows
            || !Arc::ptr_eq(mmap, down_mmap)
            || range.len() != down_stride
            || range.start != down0_range.start + i * down_stride
        {
            return None;
        }
    }

    Some(MoePackedQ4Planes::new(
        WeightBytes::Mapped {
            mmap: Arc::clone(gate_mmap),
            range: gate0_range.start..gate0_range.start + n * gate_stride,
        },
        WeightBytes::Mapped {
            mmap: Arc::clone(up_mmap),
            range: up0_range.start..up0_range.start + n * up_stride,
        },
        WeightBytes::Mapped {
            mmap: Arc::clone(down_mmap),
            range: down0_range.start..down0_range.start + n * down_stride,
        },
        gate_stride,
        up_stride,
        down_stride,
        n,
        ffn_rows,
        hidden_rows,
        gate_kind,
        up_kind,
        down_kind,
    ))
}

/// One matrix's place inside a store-backed expert's combined byte
/// buffer (gate bytes, then up, then down, concatenated by
/// `GgufExpertSource::read_expert`).
#[derive(Debug, Clone, Copy)]
pub struct StoredMatrixSpec {
    pub offset: usize,
    pub len: usize,
    pub rows: usize,
    pub cols: usize,
    pub kind: QuantKind,
}

/// Byte-range layout of one store-backed routed expert.
#[derive(Debug, Clone, Copy)]
pub struct StoredExpertLayout {
    pub gate: StoredMatrixSpec,
    pub up: StoredMatrixSpec,
    pub down: StoredMatrixSpec,
}

impl StoredExpertLayout {
    pub fn total_bytes(&self) -> usize {
        self.gate.len + self.up.len + self.down.len
    }

    /// Builds temporary zero-copy `WeightMatrix` views over a leased
    /// buffer. Each view's `WeightBytes::Shared` clone of the lease's
    /// `Arc` keeps the cache entry pinned for the view's lifetime.
    pub fn materialize(&self, lease: &ferrox_core::expert_store::ExpertLease) -> ExpertWeights {
        let mk = |spec: &StoredMatrixSpec| WeightMatrix::Quantized {
            data: WeightBytes::Shared {
                buf: lease.shared_buf(),
                range: spec.offset..spec.offset + spec.len,
            },
            rows: spec.rows,
            cols: spec.cols,
            kind: spec.kind,
        };
        ExpertWeights {
            gate: mk(&self.gate),
            up: mk(&self.up),
            down: mk(&self.down),
        }
    }
}

/// [`ExpertSource`] over a (possibly sharded) GGUF checkpoint: each
/// expert's gate/up/down byte ranges are read positionally from the
/// owning shard file and concatenated, so a store miss touches exactly
/// that expert's bytes -- no mmap of the expert region, no shared seek
/// cursor.
pub struct GgufExpertSource {
    files: Vec<std::fs::File>,
    /// (layer, expert) -> the three (file index, offset, len) segments
    /// in gate/up/down order.
    segments: std::collections::HashMap<ExpertKey, [(usize, u64, usize); 3]>,
}

impl ExpertSource for GgufExpertSource {
    fn expert_len(&self, key: ExpertKey) -> Option<usize> {
        self.segments
            .get(&key)
            .map(|segs| segs.iter().map(|&(_, _, len)| len).sum())
    }

    fn read_expert(&self, key: ExpertKey) -> std::io::Result<Vec<u8>> {
        let segs = self
            .segments
            .get(&key)
            .ok_or_else(|| std::io::Error::new(std::io::ErrorKind::NotFound, format!("{key:?}")))?;
        let total: usize = segs.iter().map(|&(_, _, len)| len).sum();
        let mut buf = vec![0u8; total];
        let mut written = 0;
        for &(fi, offset, len) in segs {
            let dst = &mut buf[written..written + len];
            #[cfg(unix)]
            {
                use std::os::unix::fs::FileExt;
                self.files[fi].read_exact_at(dst, offset)?;
            }
            #[cfg(not(unix))]
            {
                use std::io::{Read, Seek, SeekFrom};
                let mut f = &self.files[fi];
                f.seek(SeekFrom::Start(offset))?;
                f.read_exact(dst)?;
            }
            written += len;
        }
        Ok(buf)
    }
}

/// Collects the per-expert `(file, offset, len)` segments and layout
/// for one packed 3D expert tensor -- the store-backed counterpart of
/// `split_expert_tensor`, sharing its shape/offset math. Only
/// quantized dtypes are supported (an F32/BF16 expert tensor keeps the
/// resident path; the store exists for the quantized multi-hundred-GB
/// case).
/// One packed 3D expert tensor's store-backed description: the owning
/// shard index, each expert's `(offset, len)` within that shard file,
/// and the matrix spec shared by every expert's slice.
struct StoredTensorSpecs {
    shard: usize,
    per_expert: Vec<(u64, usize)>,
    spec: StoredMatrixSpec,
}

fn stored_expert_specs(
    file: &ShardedGguf,
    name: &str,
    n_experts: usize,
) -> Result<Option<StoredTensorSpecs>, LoadError> {
    let info = find_info(file, name)?;
    if info.shape.len() != 3 || info.shape[2] as usize != n_experts {
        let file_experts = info.shape.last().map(|&d| d as usize).unwrap_or(0);
        return Err(LoadError::ExpertCountMismatch(
            name.to_string(),
            file_experts,
            n_experts,
        ));
    }
    let out_dim = info.shape[1] as usize;
    let in_dim = info.shape[0] as usize;
    let Some(kind) = quant_kind_for(info.dtype) else {
        return Ok(None); // F32/BF16 (or unsupported): resident fallback
    };
    let shard = file
        .tensor_shard_index(name)
        .expect("find_info succeeded, shard index must exist");
    // The mmap range of a tensor within a GgufFile IS its byte offset
    // range within that shard file (the mmap covers the whole file).
    let (_, full_range) = file.tensor_mapped_range(name)?;
    let total_len = full_range.end - full_range.start;
    let bytes_per_expert = total_len / n_experts;
    let per_expert: Vec<(u64, usize)> = (0..n_experts)
        .map(|e| {
            (
                (full_range.start + e * bytes_per_expert) as u64,
                bytes_per_expert,
            )
        })
        .collect();
    let spec = StoredMatrixSpec {
        offset: 0, // caller assigns the position within the combined buffer
        len: bytes_per_expert,
        rows: out_dim,
        cols: in_dim,
        kind,
    };
    Ok(Some(StoredTensorSpecs {
        shard,
        per_expert,
        spec,
    }))
}

impl Decoder {
    /// Loads real weights from `path` for the given `config`. `config`
    /// supplies the architecture shape (layer count, head counts, MoE
    /// topology); tensor names are resolved against it using the
    /// llama.cpp naming convention described in the module docs.
    ///
    /// A `config.moe.n_experts <= 1` model is treated as dense: expert
    /// weights are read from the plain `blk.N.ffn_{gate,up,down}.weight`
    /// tensor names rather than the packed 3D `_exps` variant.
    pub fn from_gguf(
        path: impl AsRef<std::path::Path>,
        config: ModelConfig,
    ) -> Result<Self, LoadError> {
        Self::from_gguf_with_expert_cache(path, config, None)
    }

    /// Like `from_gguf`, but with `expert_cache_bytes: Some(budget)`
    /// routed experts are NOT loaded resident: each layer holds only
    /// byte-range layouts, and expert bytes are read on demand through
    /// one bounded, lease-protected `ExpertStore` shared by every
    /// layer (a single global byte budget; see
    /// `ferrox_core::expert_store`). Dense layers, shared experts,
    /// attention, embeddings, and the output head stay resident/mapped
    /// exactly as before -- only routed experts stream. Layers whose
    /// expert tensors are F32/BF16 fall back to resident loading (the
    /// store exists for the quantized case). Output is bit-identical
    /// to the resident path -- same bytes, same kernels -- pinned by
    /// the roundtrip suite's equivalence test.
    pub fn from_gguf_with_expert_cache(
        path: impl AsRef<std::path::Path>,
        mut config: ModelConfig,
        expert_cache_bytes: Option<u64>,
    ) -> Result<Self, LoadError> {
        let path = path.as_ref();
        let file = ShardedGguf::open(path)?;

        // gpt-oss carries five per-layer tensors the generic GQA layer
        // structs have no home for, and reuses `post_attention_norm` for
        // a *different* norm slot than Gemma does. Both are decided by
        // the architecture string, so resolve them once here. See
        // `crate::decoder::GptOssWeights` and
        // `PRE_FFN_NORM_IS_POST_ATTENTION_NORM`.
        //
        // These used to be ONE flag, `arch == "gpt-oss"`, standing for
        // two unrelated facts. Splitting them is what let `seed_oss` --
        // which shares the norm slot and has none of the extra tensors
        // -- be admitted without also being handed attention sinks.
        let arch = file
            .metadata_str("general.architecture")
            .unwrap_or_default()
            .to_string();
        let is_gpt_oss = arch == "gpt-oss";
        let post_attn_norm_is_pre_ffn_norm = pre_ffn_norm_is_post_attention_norm(&arch);
        let mut gpt_oss_layers: Vec<crate::decoder::GptOssLayer> = Vec::new();

        // One store for the whole model (keys are (layer, expert)),
        // built up-front with every stored expert's segments; created
        // only when the cache is enabled AND some layer can use it.
        let mut store_segments: std::collections::HashMap<ExpertKey, [(usize, u64, usize); 3]> =
            std::collections::HashMap::new();
        let mut stored_layouts: Vec<Option<Vec<StoredExpertLayout>>> = Vec::new();

        // Loaded like any other weight matrix: a quantized embedding
        // table stays quantized (zero-copy mmap) and token lookup
        // dequantizes one row via `WeightMatrix::dequant_row`, instead
        // of the whole vocabulary tensor being widened to f32 up front.
        let embedding = load_weight_matrix(&file, "token_embd.weight")?;

        let mut layers = Vec::with_capacity(config.n_layers);
        let mut refined_qk_norm = config.qk_norm_style;
        for l in 0..config.n_layers {
            let (q_proj, k_proj, v_proj) = load_qkv_projections(&file, l, &config)?;
            let q_norm = load_f32_vec_optional(&file, &format!("blk.{l}.attn_q_norm.weight"))?;
            let k_norm = load_f32_vec_optional(&file, &format!("blk.{l}.attn_k_norm.weight"))?;
            // Refine WholeVector vs PerHead from the first observed norm length.
            if let Some(ref w) = q_norm {
                if w.len() == config.head_dim {
                    refined_qk_norm = crate::capability::QkNormStyle::PerHead;
                } else if w.len() == config.n_heads * config.head_dim {
                    refined_qk_norm = crate::capability::QkNormStyle::WholeVector;
                } else {
                    return Err(LoadError::UnsupportedFeature(
                        config.name.to_string(),
                        format!(
                            "blk.{l}.attn_q_norm.weight length {} matches neither head_dim={} \
                             nor n_heads*head_dim={}",
                            w.len(),
                            config.head_dim,
                            config.n_heads * config.head_dim
                        ),
                    ));
                }
            }
            let attn = AttnWeights {
                q_proj,
                k_proj,
                v_proj,
                o_proj: load_weight_matrix(&file, &format!("blk.{l}.attn_output.weight"))?,
                norm_weight: load_f32_vec(&file, &format!("blk.{l}.attn_norm.weight"))?,
                q_norm,
                k_norm,
                // Qwen2/Qwen2-MoE-family real QKV bias (`attn_{q,k,v}.bias`,
                // real config `qkv_bias`, `o_proj` has none) -- see
                // `AttnWeights::q_bias`'s doc comment.
                q_bias: load_f32_vec_optional(&file, &format!("blk.{l}.attn_q.bias"))?,
                k_bias: load_f32_vec_optional(&file, &format!("blk.{l}.attn_k.bias"))?,
                v_bias: load_f32_vec_optional(&file, &format!("blk.{l}.attn_v.bias"))?,
                // gpt-oss ships `post_attention_norm` but applies it in
                // Gemma's *other* slot: llama.cpp's openai-moe graph
                // norms `ffn_inp` with it after the attention residual,
                // i.e. it is the pre-FFN norm, not a post-attention one.
                // It is read below into `MoeWeights::norm_weight`.
                post_attn_norm: if post_attn_norm_is_pre_ffn_norm {
                    None
                } else {
                    // Both spellings; see `load_norm_vec_either_spelling`.
                    load_norm_vec_either_spelling(&file, &format!("blk.{l}.post_attention_norm"))?
                },
                post_ffn_norm: load_norm_vec_either_spelling(
                    &file,
                    &format!("blk.{l}.post_ffw_norm"),
                )?,
            };

            // Leading dense layers (see ModelConfig::layer_is_dense's
            // doc comment) load from the plain dense tensor names
            // regardless of this model's global MoE topology, matching
            // the DeepSeek-2/3-family convention found in
            // ik_llama.cpp's source. A model with n_experts<=1
            // globally (the dense test fixture) is dense on every
            // layer either way.
            let is_dense_layer = config.layer_is_dense(l) || config.moe.n_experts <= 1;
            let n_experts = if is_dense_layer {
                1
            } else {
                config.moe.n_experts
            };
            let experts: ExpertBacking = if is_dense_layer {
                ExpertBacking::Resident(vec![load_dense_expert(&file, l, &config)?])
            } else {
                // Try store-backed layouts first when the cache is
                // enabled; fall back to resident when any of the three
                // tensors isn't a supported quantized dtype.
                let stored = if expert_cache_bytes.is_some() {
                    let g = stored_expert_specs(
                        &file,
                        &format!("blk.{l}.ffn_gate_exps.weight"),
                        n_experts,
                    )?;
                    let u = stored_expert_specs(
                        &file,
                        &format!("blk.{l}.ffn_up_exps.weight"),
                        n_experts,
                    )?;
                    let d = stored_expert_specs(
                        &file,
                        &format!("blk.{l}.ffn_down_exps.weight"),
                        n_experts,
                    )?;
                    match (g, u, d) {
                        (Some(gt), Some(ut), Some(dt)) => {
                            let mut layouts = Vec::with_capacity(n_experts);
                            for e in 0..n_experts {
                                let key = ExpertKey {
                                    layer: l as u32,
                                    expert: e as u32,
                                };
                                store_segments.insert(
                                    key,
                                    [
                                        (gt.shard, gt.per_expert[e].0, gt.per_expert[e].1),
                                        (ut.shard, ut.per_expert[e].0, ut.per_expert[e].1),
                                        (dt.shard, dt.per_expert[e].0, dt.per_expert[e].1),
                                    ],
                                );
                                let mut gate = gt.spec;
                                let mut up = ut.spec;
                                let mut down = dt.spec;
                                gate.offset = 0;
                                up.offset = gate.len;
                                down.offset = gate.len + up.len;
                                layouts.push(StoredExpertLayout { gate, up, down });
                            }
                            Some(layouts)
                        }
                        _ => None,
                    }
                } else {
                    None
                };
                match stored {
                    Some(layouts) => {
                        // Placeholder; the shared store is attached in a
                        // second pass below once every layer's segments
                        // are collected.
                        stored_layouts.push(Some(layouts));
                        ExpertBacking::Resident(Vec::new())
                    }
                    None => {
                        let gates = split_expert_tensor(
                            &file,
                            &format!("blk.{l}.ffn_gate_exps.weight"),
                            n_experts,
                        )?;
                        let ups = split_expert_tensor(
                            &file,
                            &format!("blk.{l}.ffn_up_exps.weight"),
                            n_experts,
                        )?;
                        let downs = split_expert_tensor(
                            &file,
                            &format!("blk.{l}.ffn_down_exps.weight"),
                            n_experts,
                        )?;
                        ExpertBacking::Resident(
                            gates
                                .into_iter()
                                .zip(ups)
                                .zip(downs)
                                .map(|((gate, up), down)| ExpertWeights { gate, up, down })
                                .collect(),
                        )
                    }
                }
            };
            if stored_layouts.len() < layers.len() + 1 {
                stored_layouts.push(None);
            }

            let shared_experts: Vec<ExpertWeights> =
                if config.moe.n_shared_experts > 0 && !is_dense_layer {
                    vec![ExpertWeights {
                        gate: load_weight_matrix(&file, &format!("blk.{l}.ffn_gate_shexp.weight"))?,
                        up: load_weight_matrix(&file, &format!("blk.{l}.ffn_up_shexp.weight"))?,
                        down: load_weight_matrix(&file, &format!("blk.{l}.ffn_down_shexp.weight"))?,
                    }]
                } else {
                    Vec::new()
                };

            let router = if !is_dense_layer {
                load_weight_matrix(&file, &format!("blk.{l}.ffn_gate_inp.weight"))?
            } else {
                // dense layer: no real router; a zero [1, hidden] matrix
                // always selects the single expert deterministically.
                WeightMatrix::F32(Tensor::zeros(vec![1, config.hidden_dim]))
            };

            let n_for_counts = match &experts {
                ExpertBacking::Resident(v) if v.is_empty() => n_experts,
                other => other.n_experts(),
            };
            let activation_counts = (0..n_for_counts)
                .map(|_| std::sync::atomic::AtomicU64::new(0))
                .collect();
            // Qwen2-MoE-specific real tensor (`blk.N.ffn_gate_inp_shexp.weight`,
            // real on-disk shape `[hidden_dim]`, confirmed against
            // llama.cpp's real `qwen2moe.cpp`) -- see
            // `MoeWeights::shared_expert_gate`'s doc comment. Presence
            // of the tensor itself is the real signal (not an
            // architecture-name list): every other supported
            // architecture's checkpoints simply don't carry this
            // tensor, so this naturally stays `None` there.
            let shared_expert_gate = if is_dense_layer {
                None
            } else {
                load_f32_vec_optional(&file, &format!("blk.{l}.ffn_gate_inp_shexp.weight"))?
            };
            #[cfg(feature = "metal")]
            let packed_q4 = match &experts {
                ExpertBacking::Resident(v) if !v.is_empty() => try_build_moe_packed_q4_planes(v),
                _ => None,
            };
            // DeepSeek-V3's aux-loss-free selection bias. The on-disk
            // name carries no `ffn_` prefix -- llama.cpp's
            // `LLM_TENSOR_FFN_EXP_PROBS_B` maps to `blk.%d.exp_probs_b`
            // (`llama-arch.cpp:416`, `gguf-py/gguf/constants.py:1240`).
            // Optional: only the DeepSeek-V3-lineage MoE recipes carry
            // it, and this same generic loader serves OLMoE / Qwen2-MoE /
            // Mixtral, which do not.
            let exp_probs_bias = if is_dense_layer {
                None
            } else {
                load_f32_vec_optional(&file, &format!("blk.{l}.exp_probs_b.bias"))?
            };
            if let Some(bias) = &exp_probs_bias {
                if bias.len() != config.moe.n_experts {
                    return Err(LoadError::UnsupportedFeature(
                        arch.clone(),
                        format!(
                            "blk.{l}.exp_probs_b.bias has {} entries but the model has {} experts",
                            bias.len(),
                            config.moe.n_experts
                        ),
                    ));
                }
                // Grouped selection masks the *biased* scores before the
                // global top-k (`build_moe_ffn`, the `n_expert_groups > 1`
                // block). ferrox's `route_top_k_grouped` takes a fixed
                // count from every group instead, which is a different
                // algorithm, so combining the two here would be a guess.
                // Refuse rather than route wrongly.
                if config.moe.expert_group_count.is_some() {
                    return Err(LoadError::UnsupportedFeature(
                        arch.clone(),
                        format!(
                            "blk.{l}.exp_probs_b.bias together with expert groups \
                             ({:?}): llama.cpp masks the biased scores per group \
                             before a global top-k, which is not the per-group \
                             top-k ferrox implements",
                            config.moe.expert_group_count
                        ),
                    ));
                }
            }
            let moe = MoeWeights {
                router,
                experts,
                shared_experts,
                shared_expert_gate,
                exp_probs_bias,
                norm_weight: if post_attn_norm_is_pre_ffn_norm {
                    // Same two spellings as above, and the same helper,
                    // so the pre-FFN-norm slot cannot drift away from
                    // the post-attention one about what a file may be
                    // called. gpt-oss and seed_oss both write `.weight`
                    // today; sharing the rule is what stops that being
                    // a thing to rediscover.
                    load_norm_vec_either_spelling(&file, &format!("blk.{l}.post_attention_norm"))?
                        .ok_or_else(|| {
                        LoadError::Gguf(GgufError::TensorNotFound(format!(
                            "blk.{l}.post_attention_norm[.weight]"
                        )))
                    })?
                } else {
                    load_f32_vec(&file, &format!("blk.{l}.ffn_norm.weight"))?
                },
                activation_counts,
                #[cfg(feature = "metal")]
                packed_q4,
            };

            if is_gpt_oss {
                gpt_oss_layers.push(load_gpt_oss_layer(&file, l, &config)?);
            }

            layers.push(LayerWeights { attn, moe });
        }

        let final_norm = load_f32_vec(&file, "output_norm.weight")?;
        // Many small Llama/Gemma-family GGUFs tie the lm-head to
        // `token_embd.weight` and omit `output.weight` (llama.cpp
        // `llama_model_loader` falls back the same way). Prefer the
        // explicit head when present.
        let output_head = match load_weight_matrix(&file, "output.weight") {
            Ok(w) => w,
            Err(_) => load_weight_matrix(&file, "token_embd.weight")?,
        };

        // Second pass: attach the one shared store to every
        // store-backed layer. Opening the shard files fresh (plain
        // `File` handles for positional reads, not mmaps) keeps the
        // stored experts' bytes out of the process's mapped footprint
        // entirely.
        if !store_segments.is_empty() {
            let budget = expert_cache_bytes
                .expect("store_segments only populated when a cache budget is set")
                as usize;
            let files: Result<Vec<std::fs::File>, std::io::Error> =
                file.shard_paths().iter().map(std::fs::File::open).collect();
            let files = files.map_err(GgufError::from)?;
            let store = std::sync::Arc::new(ExpertStore::new(
                GgufExpertSource {
                    files,
                    segments: store_segments,
                },
                budget,
            ));
            for (l, layer) in layers.iter_mut().enumerate() {
                if let Some(layouts) = stored_layouts.get_mut(l).and_then(Option::take) {
                    layer.moe.experts = ExpertBacking::Stored {
                        store: std::sync::Arc::clone(&store),
                        layouts,
                        layer: l as u32,
                    };
                }
            }
        }

        config.qk_norm_style = refined_qk_norm;

        let family = crate::capability::resolve_profile(
            file.metadata_str("general.architecture").unwrap_or("llama"),
        )
        .map(|p| p.family)
        .unwrap_or(crate::capability::DecoderFamily::StandardGqa);
        let memory_kind = crate::capability::resolve_profile(
            file.metadata_str("general.architecture").unwrap_or("llama"),
        )
        .map(|p| p.memory)
        .unwrap_or(crate::capability::MemoryKind::KvGqa);
        let execution_plan = crate::execution_plan::ExecutionPlan::from_config(
            &config,
            family,
            memory_kind,
            crate::execution_plan::ExecutionPlan::probe_metal_caps(),
        );

        let decoder = Decoder {
            config,
            embedding,
            layers,
            final_norm,
            output_head,
            gpu_vram_budget_bytes: None,
            gpt_oss: if is_gpt_oss {
                Some(crate::decoder::GptOssWeights {
                    layers: gpt_oss_layers,
                })
            } else {
                None
            },
            qk_norm_after_rope: QK_NORM_AFTER_ROPE_ARCHITECTURES.contains(&arch.as_str()),
            #[cfg(feature = "metal")]
            metal_attn_kv: std::sync::Mutex::new(None),
            execution_plan,
            kv_window: crate::decoder::KvWindowPolicy::from_env(),
            plan_cache: std::sync::Mutex::new(std::collections::HashMap::new()),
        };
        // Resolve every kernel the model will need while we still have a
        // load-time error path to report it on, then seal: from here a
        // lookup that misses is an unpredicted slow path and says so.
        decoder.probe_kernels();
        ferrox_core::kernel_registry::seal_or_error()
            .map_err(|e| LoadError::StrictKernels(e.to_string()))?;
        // `ModelConfig` is parsed from a *different* handle on the same
        // file (the CLI opens its own `GgufFile`, then hands the config
        // here), so the model-level tensors it consumed were recorded on
        // that handle, not this one. Replay them before the gate, or
        // every Llama-3.x checkpoint reads as carrying an unread
        // `rope_freqs.weight` it in fact uses on every RoPE call.
        for name in crate::config::MODEL_LEVEL_TENSORS_READ_BY_CONFIG {
            file.note_consumed(name);
        }
        assert_every_tensor_consumed(&file)?;
        Ok(decoder)
    }
}

/// Tensor-name prefixes a text-generation load legitimately never
/// reads. Everything here is consumed by a *different* code path, not by
/// nothing: multimodal projector planes belong to `mmproj`, and the
/// per-shard split bookkeeping is metadata, not weights.
const IGNORED_TENSOR_PREFIXES: &[&str] = &["mm.", "v.", "mmproj.", "resampler.", "audio."];

/// Fails the load when the checkpoint carries tensors this build never
/// looked at.
///
/// A tensor nobody reads is not a harmless extra: it is a term of the
/// real graph that ours is missing. gpt-oss ships `blk.N.attn_sinks`
/// and ferrox has no attention-sink code anywhere, so the file loads,
/// runs at full speed, and emits a different distribution than the model
/// it claims to be; the newer MoE recipes ship `ffn_exp_probs_b` the
/// same way. Both are silent today, and both are exactly what the
/// architecture registry cannot catch, because the architecture *string*
/// is one ferrox does support -- it is the checkpoint that carries more
/// than the registry entry promises.
///
/// This is deliberately the last check in the load: by here every loader
/// arm has had its chance to ask for what it needs, so what is left over
/// is what nothing in this build knows about.
///
/// `FERROX_ALLOW_UNKNOWN_TENSORS=1` downgrades it to a warning, for the
/// case where a human has decided the missing term does not matter (a
/// bias tensor of zeros, an auxiliary head that never runs). The default
/// is refusal: a wrong answer is worse than no answer.
pub fn assert_every_tensor_consumed(file: &ShardedGguf) -> Result<(), LoadError> {
    let mut left: Vec<String> = file
        .unconsumed_tensors()
        .into_iter()
        .filter(|n| !IGNORED_TENSOR_PREFIXES.iter().any(|p| n.starts_with(p)))
        .collect();
    if left.is_empty() {
        return Ok(());
    }
    left.sort();
    let shown = left.iter().take(8).cloned().collect::<Vec<_>>().join(", ");
    let listing = if left.len() > 8 {
        format!("{shown}, … (+{} more)", left.len() - 8)
    } else {
        shown
    };
    if matches!(
        std::env::var("FERROX_ALLOW_UNKNOWN_TENSORS")
            .ok()
            .as_deref(),
        Some("1") | Some("true") | Some("on")
    ) {
        eprintln!(
            "ferrox: WARNING -- {} tensor(s) in this checkpoint are never read \
             ({listing}); output may be wrong (FERROX_ALLOW_UNKNOWN_TENSORS=1)",
            left.len()
        );
        return Ok(());
    }
    Err(LoadError::UnconsumedTensors(left.len(), listing))
}

#[cfg(test)]
mod tests {

    /// A quantized 1-D tensor loads through the shared helper.
    ///
    /// This used to be six copies of `load_f32_vec`, and they had
    /// drifted badly: this one decoded twenty dtypes while the five
    /// architecture loaders decoded three (F32/F16/BF16). A quantizer
    /// that emits a Q8_0 norm or bias -- ordinary for aggressive
    /// quants -- loaded on the generic path and was rejected with
    /// `UnsupportedDtype` on GLM-5.2, Kimi, DeepSeek-MLA, Gemma-4 and
    /// the hybrid stack.
    ///
    /// This file's own comment predicted exactly that, about the same
    /// split one level down: "a dtype ferrox can decode should never be
    /// rejected here just because the *other* dispatch table below
    /// knows it -- that split is how a supported format turns into a
    /// load failure on the one checkpoint that uses it."
    #[test]
    fn a_quantized_one_dimensional_tensor_widens_through_the_shared_helper() {
        let values: Vec<f32> = (0..64).map(|i| (i as f32 - 32.0) * 0.25).collect();
        let quantized = ferrox_quant::quantize_q8_0(&values);

        struct OneTensor {
            info: TensorInfo,
            bytes: Vec<u8>,
        }
        impl TensorSource for OneTensor {
            fn metadata(&self, _key: &str) -> Option<&ferrox_gguf::GgufValue> {
                None
            }
            fn find_tensor(&self, name: &str) -> Option<&TensorInfo> {
                (name == self.info.name).then_some(&self.info)
            }
            fn tensor_bytes(&self, _name: &str) -> Result<&[u8], GgufError> {
                Ok(&self.bytes)
            }
            fn tensor_mapped_range(
                &self,
                name: &str,
            ) -> Result<
                (
                    std::sync::Arc<ferrox_gguf::MmapHandle>,
                    std::ops::Range<usize>,
                ),
                GgufError,
            > {
                // Never reached: `load_f32_vec` widens from bytes.
                Err(GgufError::TensorNotFound(name.to_string()))
            }
        }

        let source = OneTensor {
            info: TensorInfo {
                name: "blk.0.attn_norm.weight".to_string(),
                shape: vec![64],
                dtype: GgmlType::Q8_0,
                offset: 0,
            },
            bytes: quantized,
        };

        let widened = load_f32_vec(&source, "blk.0.attn_norm.weight")
            .expect("a Q8_0 norm must load, not report an unsupported dtype");
        assert_eq!(widened.len(), values.len());
        for (got, want) in widened.iter().zip(values.iter()) {
            assert!(
                (got - want).abs() < 0.05,
                "q8_0 round trip: got {got}, want {want}"
            );
        }
    }
    use super::*;
    use byteorder::{LittleEndian, WriteBytesExt};
    use std::io::Write;

    fn write_string(buf: &mut Vec<u8>, s: &str) {
        buf.write_u64::<LittleEndian>(s.len() as u64).unwrap();
        buf.write_all(s.as_bytes()).unwrap();
    }

    fn write_kv_str(buf: &mut Vec<u8>, key: &str, val: &str) {
        write_string(buf, key);
        buf.write_u32::<LittleEndian>(8).unwrap(); // type = string
        write_string(buf, val);
    }

    /// A minimal, tensor-free GGUF byte buffer declaring only
    /// `general.architecture` (no `{arch}.block_count` or any other
    /// hparam key) -- the shape a stripped-down or malformed file might
    /// take, and the exact case `ModelConfig::from_gguf` must reject
    /// loudly rather than silently default around.
    fn build_arch_only_gguf(arch: &str) -> Vec<u8> {
        let mut buf = Vec::new();
        buf.write_u32::<LittleEndian>(ferrox_gguf::GGUF_MAGIC)
            .unwrap();
        buf.write_u32::<LittleEndian>(3).unwrap(); // version
        buf.write_u64::<LittleEndian>(0).unwrap(); // tensor_count
        buf.write_u64::<LittleEndian>(1).unwrap(); // kv_count
        write_kv_str(&mut buf, "general.architecture", arch);
        buf
    }

    #[test]
    fn model_config_from_gguf_fails_loudly_when_required_hparams_are_missing() {
        let tmp =
            std::env::temp_dir().join(format!("ferrox_test_arch_only_{}.gguf", std::process::id()));
        // Use a registered architecture so the failure is MissingHparam,
        // not UnsupportedArchitecture.
        std::fs::write(&tmp, build_arch_only_gguf("llama")).unwrap();
        let file = ferrox_gguf::GgufFile::open(&tmp).expect("minimal header must still parse");
        std::fs::remove_file(&tmp).ok();

        match ModelConfig::from_gguf(&file) {
            Err(LoadError::MissingHparam(key)) => {
                assert_eq!(key, "llama.block_count");
            }
            other => panic!(
                "expected LoadError::MissingHparam for a file with no hparam keys, got {other:?}"
            ),
        }
    }

    #[test]
    fn model_config_from_gguf_fails_closed_on_unknown_architecture() {
        let tmp = std::env::temp_dir().join(format!(
            "ferrox_test_unknown_arch_{}.gguf",
            std::process::id()
        ));
        std::fs::write(&tmp, build_arch_only_gguf("bogus-arch-with-no-hparams")).unwrap();
        let file = ferrox_gguf::GgufFile::open(&tmp).expect("minimal header must still parse");
        std::fs::remove_file(&tmp).ok();

        match ModelConfig::from_gguf(&file) {
            Err(LoadError::UnsupportedArchitecture(arch)) => {
                assert_eq!(arch, "bogus-arch-with-no-hparams");
            }
            other => panic!(
                "expected LoadError::UnsupportedArchitecture for an unregistered arch, got {other:?}"
            ),
        }
    }

    fn write_kv_f32(buf: &mut Vec<u8>, key: &str, val: f32) {
        write_string(buf, key);
        buf.write_u32::<LittleEndian>(6).unwrap(); // type = float32
        buf.write_f32::<LittleEndian>(val).unwrap();
    }

    /// `arch` plus one f32 hparam, so a metadata-only feature gate can be
    /// exercised without building a whole checkpoint.
    fn build_arch_plus_f32_gguf(arch: &str, key: &str, val: f32) -> Vec<u8> {
        let mut buf = Vec::new();
        buf.write_u32::<LittleEndian>(ferrox_gguf::GGUF_MAGIC)
            .unwrap();
        buf.write_u32::<LittleEndian>(3).unwrap(); // version
        buf.write_u64::<LittleEndian>(0).unwrap(); // tensor_count
        buf.write_u64::<LittleEndian>(2).unwrap(); // kv_count
        write_kv_str(&mut buf, "general.architecture", arch);
        write_kv_f32(&mut buf, key, val);
        buf
    }

    fn config_error_for(arch: &str, key: &str, val: f32, tag: &str) -> LoadError {
        let tmp = std::env::temp_dir().join(format!("ferrox_test_scale_{tag}.gguf"));
        std::fs::write(&tmp, build_arch_plus_f32_gguf(arch, key, val)).unwrap();
        let file = ferrox_gguf::GgufFile::open(&tmp).expect("minimal header must still parse");
        std::fs::remove_file(&tmp).ok();
        ModelConfig::from_gguf(&file).expect_err("must not succeed")
    }

    /// Granite / MiniCPM / Command-R multipliers are hparams, not
    /// tensors, so `assert_every_tensor_consumed` cannot see them: a
    /// checkpoint declaring one loads, runs at full speed, and computes
    /// a differently-scaled graph than it was trained as. Refuse by name
    /// until the math lands.
    #[test]
    fn a_declared_multiplier_this_decoder_does_not_apply_is_refused_by_name() {
        for (key, val) in [
            ("granite.logit_scale", 6.0f32),
            ("granite.residual_scale", 0.22),
            ("granite.embedding_scale", 12.0),
            ("granite.attention.scale", 0.015_625),
        ] {
            let tag = key.replace('.', "_");
            match config_error_for("granite", key, val, &tag) {
                LoadError::UnsupportedFeature(arch, msg) => {
                    assert_eq!(arch, "granite");
                    assert!(msg.contains(key), "error must name the key: {msg}");
                }
                other => panic!("expected UnsupportedFeature for {key}, got {other:?}"),
            }
        }
    }

    /// The gate must not fire on a multiplier that is a no-op. A file
    /// writing `residual_scale = 1.0` describes the graph ferrox already
    /// computes, and refusing it would be a false alarm. llama.cpp's
    /// `f_attention_scale` uses `0.0` rather than `1.0` as its "unset"
    /// sentinel, so the two are checked against their own no-op values.
    #[test]
    fn a_multiplier_that_is_a_no_op_is_not_refused() {
        for (key, val) in [
            ("granite.logit_scale", 1.0f32),
            ("granite.residual_scale", 1.0),
            ("granite.embedding_scale", 1.0),
            ("granite.attention.scale", 0.0),
        ] {
            let tag = format!("noop_{}", key.replace('.', "_"));
            // The file carries no `block_count`, so the load still fails
            // -- but on the *missing hparam*, having passed this gate.
            match config_error_for("granite", key, val, &tag) {
                LoadError::MissingHparam(k) => assert_eq!(k, "granite.block_count"),
                other => panic!("no-op {key}={val} must pass the scaling gate, got {other:?}"),
            }
        }
    }

    /// One GGUF metadata value, in the three types these header-only
    /// fixtures need.
    enum Kv<'a> {
        Str(&'a str),
        U32(u32),
        F32(f32),
        /// A uint32 ARRAY. Only one gate needs it -- the sliding-window
        /// pattern, which llama.cpp reads with `ml.get_key_or_arr` --
        /// and without it that gate could only be tested through a
        /// value of some other type, which is not the case it exists
        /// for.
        Arr32(&'a [u32]),
    }

    /// A tensor-free GGUF carrying exactly `kvs` -- enough for
    /// `ModelConfig::from_gguf` to run without a single weight on disk.
    fn build_metadata_gguf(kvs: &[(&str, Kv)]) -> Vec<u8> {
        let mut buf = Vec::new();
        buf.write_u32::<LittleEndian>(ferrox_gguf::GGUF_MAGIC)
            .unwrap();
        buf.write_u32::<LittleEndian>(3).unwrap(); // version
        buf.write_u64::<LittleEndian>(0).unwrap(); // tensor_count
        buf.write_u64::<LittleEndian>(kvs.len() as u64).unwrap();
        for (k, v) in kvs {
            match v {
                Kv::Str(s) => write_kv_str(&mut buf, k, s),
                Kv::U32(n) => {
                    write_string(&mut buf, k);
                    buf.write_u32::<LittleEndian>(4).unwrap(); // type = uint32
                    buf.write_u32::<LittleEndian>(*n).unwrap();
                }
                Kv::F32(f) => write_kv_f32(&mut buf, k, *f),
                Kv::Arr32(values) => {
                    write_string(&mut buf, k);
                    buf.write_u32::<LittleEndian>(9).unwrap(); // type = array
                    buf.write_u32::<LittleEndian>(4).unwrap(); // element type = uint32
                    buf.write_u64::<LittleEndian>(values.len() as u64).unwrap();
                    for v in *values {
                        buf.write_u32::<LittleEndian>(*v).unwrap();
                    }
                }
            }
        }
        buf
    }

    fn open_metadata_gguf(tag: &str, kvs: &[(&str, Kv)]) -> ferrox_gguf::GgufFile {
        let tmp = std::env::temp_dir().join(format!("ferrox_test_meta_{tag}.gguf"));
        std::fs::write(&tmp, build_metadata_gguf(kvs)).unwrap();
        let file = ferrox_gguf::GgufFile::open(&tmp).expect("header-only file must parse");
        std::fs::remove_file(&tmp).ok();
        file
    }

    /// A minimal `llama` hparam set (64-wide single head, base 10000)
    /// plus whatever RoPE-scaling keys a test wants to add.
    fn llama_config_with(tag: &str, extra: &[(&str, Kv)]) -> ModelConfig {
        let mut kvs: Vec<(&str, Kv)> = vec![
            ("general.architecture", Kv::Str("llama")),
            ("llama.block_count", Kv::U32(1)),
            ("llama.embedding_length", Kv::U32(64)),
            ("llama.attention.head_count", Kv::U32(1)),
            ("llama.attention.head_count_kv", Kv::U32(1)),
            ("llama.attention.key_length", Kv::U32(64)),
            ("llama.rope.freq_base", Kv::F32(10_000.0)),
        ];
        for (k, v) in extra {
            kvs.push((
                k,
                match v {
                    Kv::Str(s) => Kv::Str(s),
                    Kv::U32(n) => Kv::U32(*n),
                    Kv::F32(f) => Kv::F32(*f),
                    Kv::Arr32(a) => Kv::Arr32(a),
                },
            ));
        }
        ModelConfig::from_gguf(&open_metadata_gguf(tag, &kvs)).expect("fixture must load")
    }

    /// Builds a config for an arbitrary architecture tag, returning the
    /// error rather than unwrapping it.
    /// llama.cpp chooses the FFN gate activation PER ARCHITECTURE;
    /// ferrox chose it per family. Those are different partitions, and
    /// `grok` is where they disagree: `src/models/grok.cpp:165` passes
    /// `LLM_FFN_GELU` to `build_moe_ffn`, while `grok` is
    /// `DecoderFamily::StandardGqa` and so was handed SwiGLU -- a
    /// different FFN on every layer.
    ///
    /// Latent, because `grok` is not audited and refuses today. Pinned
    /// anyway: the failure mode is that auditing it later makes it
    /// silently wrong, and an audit is exactly when nobody thinks to
    /// re-check the activation.
    #[test]
    fn the_ffn_activation_follows_the_architecture_not_the_family() {
        use crate::capability::uses_geglu;
        use crate::config::FfnActivation;

        assert!(uses_geglu("grok"), "grok's MoE FFN gate is GELU upstream");
        // Same family, SiLU upstream (`src/models/dbrx.cpp:122`), so the
        // family rule alone cannot be what selects grok.
        assert!(!uses_geglu("dbrx"));
        assert!(!uses_geglu("llama"));

        // The Gemma lineage keeps its GELU through the FAMILY rule, so
        // the new per-architecture arm must not have displaced it.
        // gemma2/gemma3 only: `gemma` v1 is unaudited and refuses, so
        // it cannot be loaded to check its activation.
        for gemma in ["gemma2", "gemma3"] {
            assert!(
                !uses_geglu(gemma),
                "{gemma} is GELU via GemmaFamily; listing it here too \
                 would hide a later regression in the family rule"
            );
            assert_eq!(
                config_for_arch(gemma).expect("gemma loads").ffn_activation,
                FfnActivation::Gelu,
                "{gemma}"
            );
        }

        // And a plain SwiGLU architecture stays SwiGLU.
        assert_eq!(
            config_for_arch("llama")
                .expect("llama loads")
                .ffn_activation,
            FfnActivation::Swiglu
        );
    }

    /// The no-renormalise list is keyed on what llama.cpp's GRAPH does,
    /// not on what a GGUF says, because for these architectures the
    /// GGUF says nothing.
    ///
    /// `expert_weights_norm` is only written by converters that set it.
    /// `deepseek.cpp:145` passes `norm_w=false`, and
    /// `conversion/deepseek.py`'s `DeepseekModel` never writes the key
    /// -- only `DeepseekV2Model` does. So a real `deepseek` checkpoint
    /// carries no key at all and ferrox fell through to its default,
    /// renormalising the selected experts' softmax weights where
    /// llama.cpp leaves them alone.
    ///
    /// The same mistake made OLMoE emit garbage, which is why that list
    /// exists. This pins the membership so a later edit cannot quietly
    /// drop a name back into the renormalising default.
    #[test]
    fn the_architectures_llama_cpp_does_not_renormalise_are_pinned() {
        for arch in ["deepseek", "olmoe", "qwen2moe"] {
            assert!(
                NO_TOPK_RENORMALIZE_ARCHITECTURES.contains(&arch),
                "{arch} passes norm_w=false in llama.cpp and must not be renormalised"
            );
        }
        // `deepseek2` is a DIFFERENT architecture whose converter DOES
        // write the key, so it must not be on this list -- it gets its
        // answer from the file.
        assert!(!NO_TOPK_RENORMALIZE_ARCHITECTURES.contains(&"deepseek2"));
        assert!(!NO_TOPK_RENORMALIZE_ARCHITECTURES.contains(&"qwen3moe"));
    }

    /// Every architecture llama.cpp defaults to SIGMOID gating must be
    /// on the list, because for these the GGUF carries no key to say so.
    ///
    /// Each of these reads `LLM_KV_EXPERT_GATING_FUNC` as optional and
    /// then sets SIGMOID when it is absent, so a converted checkpoint
    /// has nothing in it that would correct ferrox's softmax default.
    /// Same shape as the `deepseek` top-k renormalisation bug, and as
    /// `phi3`'s sliding window: the file is silent and the architecture
    /// decides.
    #[test]
    fn the_architectures_llama_cpp_defaults_to_sigmoid_gating_are_pinned() {
        for arch in ["afmoe", "deepseek2", "glm4moe", "laguna", "step35"] {
            assert!(
                SIGMOID_GATING_ARCHITECTURES.contains(&arch),
                "{arch} sets SIGMOID when the gating key is absent"
            );
        }
        // Architectures that HARDCODE softmax must stay off it, or the
        // fix becomes the opposite bug: `ernie4-5-moe.cpp:90` and
        // `qwen3moe` both gate with softmax unconditionally.
        for softmax in ["ernie4_5-moe", "qwen3moe", "olmoe", "llama"] {
            assert!(
                !SIGMOID_GATING_ARCHITECTURES.contains(&softmax),
                "{softmax} does not default to sigmoid"
            );
        }
    }

    /// Every name in every architecture-keyed behaviour table is a name
    /// the catalog actually resolves, on the generic-GQA path.
    ///
    /// These five tables are the repo's dominant bug shape in its purest
    /// form: five lists of strings that have to agree with a sixth
    /// structure (`capability::architecture_catalog`) about what an
    /// architecture is called, with nothing checking it. A typo, a
    /// hyphen where the GGUF has an underscore, or a name that later
    /// moves to a dedicated stack all produce the same thing -- an entry
    /// that reads as coverage and can never fire. This repo has shipped
    /// exactly that once already, in `unsupported_feature_keys`, keyed
    /// on a GGUF spelling no converter writes.
    ///
    /// The generic-path check is the second half and the sharper one: a
    /// behaviour flag on an architecture that is `DedicatedOnly` or
    /// `Deferred` never reaches this loader, so it is dead text.
    ///
    /// Sabotage to confirm: add `"seedoss"` to any list below.
    #[test]
    fn every_architecture_keyed_behaviour_table_names_a_real_generic_row() {
        let tables: &[(&str, &[&str])] = &[
            ("SIGMOID_GATING_ARCHITECTURES", SIGMOID_GATING_ARCHITECTURES),
            (
                "NO_TOPK_RENORMALIZE_ARCHITECTURES",
                NO_TOPK_RENORMALIZE_ARCHITECTURES,
            ),
            (
                "PRE_FFN_NORM_IS_POST_ATTENTION_NORM",
                PRE_FFN_NORM_IS_POST_ATTENTION_NORM,
            ),
            ("LEADING_DENSE_KEY_IS_INERT", LEADING_DENSE_KEY_IS_INERT),
            (
                "QK_NORM_AFTER_ROPE_ARCHITECTURES",
                QK_NORM_AFTER_ROPE_ARCHITECTURES,
            ),
        ];
        for (table, names) in tables {
            for arch in *names {
                let profile = crate::capability::resolve_profile(arch).unwrap_or_else(|| {
                    panic!("{table} names `{arch}`, which the catalog does not have")
                });
                if matches!(profile.path, crate::capability::ArchPath::GenericGqa { .. }) {
                    continue;
                }
                // Not a generic row, so the entry cannot fire HERE.
                // That is allowed only when something else is named as
                // applying the behaviour instead. An unexplained dead
                // entry still fails, which is the whole point.
                let owner = DEDICATED_OWNS_ITS_BEHAVIOUR
                    .iter()
                    .find(|(name, _)| name == arch)
                    .map(|(_, owner)| *owner);
                assert!(
                    owner.is_some(),
                    "{table} names `{arch}`, which resolves to {:?} and never reaches this \
                     loader, so the entry cannot fire. Either drop it, or add it to \
                     DEDICATED_OWNS_ITS_BEHAVIOUR naming what applies the behaviour instead",
                    profile.path
                );
            }
        }
    }

    /// The three tables that describe how a layer is BUILT, rather than
    /// how it is routed, only carry architectures that are audited.
    ///
    /// The distinction matters and is not pedantry. A routing default
    /// (`SIGMOID_GATING_ARCHITECTURES`, `NO_TOPK_RENORMALIZE_ARCHITECTURES`)
    /// is allowed to name an architecture that still refuses: it is
    /// written down ahead of time so a later admission inherits the
    /// right answer, and the tables say so. But the three below change
    /// which TENSOR a layer reads and in what order -- and each was
    /// added for exactly one architecture, whose fixture is the only
    /// thing proving the change is right. A fourth name appearing here
    /// without evidence would be a claim about a graph nobody read,
    /// carried by a list whose doc comment cites two.
    #[test]
    fn the_layer_shape_tables_only_name_audited_architectures() {
        for (table, names) in [
            (
                "PRE_FFN_NORM_IS_POST_ATTENTION_NORM",
                PRE_FFN_NORM_IS_POST_ATTENTION_NORM,
            ),
            ("LEADING_DENSE_KEY_IS_INERT", LEADING_DENSE_KEY_IS_INERT),
            (
                "QK_NORM_AFTER_ROPE_ARCHITECTURES",
                QK_NORM_AFTER_ROPE_ARCHITECTURES,
            ),
        ] {
            for arch in names {
                assert!(
                    crate::capability::is_audited_generic(arch),
                    "{table} names `{arch}`, which is not in AUDITED_GENERIC_GQA. Either it \
                     has a fixture proving the change is right -- audit it -- or the entry \
                     is a guess about a graph"
                );
            }
        }
    }

    fn config_for_arch(arch: &'static str) -> Result<ModelConfig, LoadError> {
        // The per-arch hyperparameter keys are looked up by the arch's
        // own prefix, so they have to be built for the arch under test.
        let keys: Vec<String> = [
            "block_count",
            "embedding_length",
            "attention.head_count",
            "attention.head_count_kv",
            "attention.key_length",
        ]
        .iter()
        .map(|k| format!("{arch}.{k}"))
        .collect();
        let theta = format!("{arch}.rope.freq_base");
        let kvs: Vec<(&str, Kv)> = vec![
            ("general.architecture", Kv::Str(arch)),
            (keys[0].as_str(), Kv::U32(1)),
            (keys[1].as_str(), Kv::U32(64)),
            (keys[2].as_str(), Kv::U32(1)),
            (keys[3].as_str(), Kv::U32(1)),
            (keys[4].as_str(), Kv::U32(64)),
            (theta.as_str(), Kv::F32(10_000.0)),
        ];
        ModelConfig::from_gguf(&open_metadata_gguf(arch, &kvs))
    }

    /// The generic path is OPT-IN, and this is what proves it.
    ///
    /// An architecture nobody has checked used to FALL ONTO generic GQA
    /// and run. Five did exactly that and computed the wrong thing for
    /// the life of the project. The refusal exists; nothing tested it,
    /// so a reordering or an unevidenced addition to
    /// `AUDITED_GENERIC_GQA` would have gone unnoticed.
    #[test]
    fn an_unaudited_generic_architecture_refuses_rather_than_guessing() {
        // `nanbeige` is on the generic path and is not in the audited
        // list. It is the third name to hold this slot: `starcoder` was
        // first, until an audit found it REQUIRES a fused
        // `attn_qkv.bias` and a learned `position_embd` the generic
        // decoder has no slot for, so it refuses for a stronger reason;
        // then `xverse`, until it was admitted with a libllama-golden
        // fixture (`tests/fixture_away_graphs.rs`). `nanbeige` cannot go
        // the same way soon: it runs the same physical layers more than
        // once (`src/models/nanbeige.cpp:13-31`), which is NEW CODE, and
        // its blocker is invisible in metadata, so nothing but this gate
        // stops it.
        assert!(
            !crate::capability::is_audited_generic("nanbeige"),
            "this test needs an arch that is generic AND unaudited"
        );
        match config_for_arch("nanbeige") {
            Err(LoadError::UnauditedArchitecture(name, ..)) => assert_eq!(name, "nanbeige"),
            other => panic!("expected an unaudited refusal, got {other:?}"),
        }
    }

    /// An architecture with evidence still loads, or the inversion would
    /// have turned every model off.
    #[test]
    fn an_audited_architecture_still_loads() {
        assert!(crate::capability::is_audited_generic("llama"));
        assert!(config_for_arch("llama").is_ok());
    }

    /// A NAMED problem must outrank "unaudited".
    ///
    /// `gpt2` uses learned absolute position embeddings, and that is
    /// what its refusal should say. Reporting "unaudited" instead would
    /// be true and far less useful, and it is the ordering the loader's
    /// own comment claims. Nothing checked that claim.
    #[test]
    fn a_named_refusal_outranks_the_unaudited_one() {
        let err = config_for_arch("gpt2").expect_err("gpt2 must refuse");
        assert!(
            !matches!(err, LoadError::UnauditedArchitecture(..)),
            "gpt2 should report its own reason, not that nobody audited it: {err:?}"
        );
    }

    /// A checkpoint that declares YaRN gets the per-band divisors the
    /// reference's `"yarn"` arm implies, folded into `rope_freqs` so the
    /// existing RoPE kernels apply them. Expected values are hand-derived
    /// from `_find_correction_dim` for this fixture (rotary width 64,
    /// base 10000, original context 131072): `low = 22`, `high = 35`.
    ///
    /// Before this, ferrox read neither `rope.scaling.type` nor
    /// `rope.scaling.factor`, so this file roped exactly like an
    /// unscaled one -- correct near position 0, progressively wrong
    /// further in.
    #[test]
    fn a_gguf_declaring_yarn_gets_its_rope_frequencies_rewritten() {
        let cfg = llama_config_with(
            "yarn",
            &[
                ("llama.rope.scaling.type", Kv::Str("yarn")),
                ("llama.rope.scaling.factor", Kv::F32(8.0)),
                (
                    "llama.rope.scaling.original_context_length",
                    Kv::U32(131_072),
                ),
            ],
        );
        let factors = cfg
            .rope_freqs
            .expect("a YaRN checkpoint must carry rewritten per-band frequencies")
            .full;
        assert_eq!(factors.len(), 32, "one divisor per rotation band");
        assert!(
            (factors[0] - 1.0).abs() < 1e-6,
            "the fastest band is left extrapolated, got {}",
            factors[0]
        );
        let ramp = (31.0 - 22.0) / (35.0 - 22.0);
        let want = 1.0 / (ramp / 8.0 + (1.0 - ramp));
        assert!(
            (factors[31] - want).abs() < 1e-4,
            "slowest band: got {}, reference {want}",
            factors[31]
        );
    }

    /// The rewrite must not fire on a file that did not ask for it. A
    /// scaling type ferrox does not implement (`linear`, `longrope`) is
    /// left exactly as it was rather than being roped as YaRN, which
    /// would be a new kind of wrong rather than the current known one.
    /// `rope.scaling.type = "linear"` must actually scale.
    ///
    /// Rotating position `p/s` is the same as rotating `p` with every
    /// band's frequency divided by `s`, and `rope_freqs` is exactly a
    /// per-band frequency divisor, so a uniform vector of `s` expresses
    /// linear scaling with no new code on the RoPE paths.
    ///
    /// Before this, the scaling type was compared against "yarn" and
    /// anything else returned None, so such a file loaded and roped at
    /// unscaled positions: a different model, no error.
    #[test]
    fn linear_scaling_is_applied_as_a_uniform_frequency_divisor() {
        let cfg = llama_config_with(
            "linear",
            &[
                ("llama.rope.scaling.type", Kv::Str("linear")),
                ("llama.rope.scaling.factor", Kv::F32(4.0)),
            ],
        );
        let freqs = &cfg
            .rope_freqs
            .as_ref()
            .expect("linear scaling must produce frequency factors")
            .full;
        assert_eq!(freqs.len(), cfg.head_dim / 2, "one factor per rotated pair");
        assert!(
            freqs.iter().all(|f| (*f - 4.0).abs() < 1e-6),
            "linear scaling is uniform across bands, unlike YaRN: got {freqs:?}"
        );
    }

    /// A factor that corrects nothing is not a correction.
    #[test]
    fn a_linear_factor_of_one_is_treated_as_absent() {
        assert!(llama_config_with(
            "linear_one",
            &[
                ("llama.rope.scaling.type", Kv::Str("linear")),
                ("llama.rope.scaling.factor", Kv::F32(1.0)),
            ],
        )
        .rope_freqs
        .is_none());
    }

    #[test]
    fn a_gguf_without_yarn_scaling_keeps_its_rope_frequencies_untouched() {
        assert!(llama_config_with("noscale", &[]).rope_freqs.is_none());
        // Linear scaling is NOT "no scaling". It used to land here,
        // asserted as `is_none()`, on the reasoning that leaving
        // positions alone beat roping them wrong in a new way. Both are
        // wrong output: llama.cpp divides the positions by the factor.
        // See `linear_scaling_is_applied_as_a_uniform_frequency_divisor`.
        // YaRN with a no-op factor is not a correction either.
        assert!(llama_config_with(
            "yarn_factor_one",
            &[
                ("llama.rope.scaling.type", Kv::Str("yarn")),
                ("llama.rope.scaling.factor", Kv::F32(1.0)),
                (
                    "llama.rope.scaling.original_context_length",
                    Kv::U32(131_072),
                ),
            ],
        )
        .rope_freqs
        .is_none());
    }

    /// The correction range is measured against the context the
    /// checkpoint was *trained* at, so a file that declares YaRN without
    /// `rope.scaling.original_context_length` leaves the rotation alone
    /// rather than inventing a trained length (`context_length` on such
    /// a file is the *extended* one, which would put the ramp in the
    /// wrong place at every band).
    #[test]
    fn yarn_without_an_original_context_length_is_not_guessed_at() {
        let cfg = llama_config_with(
            "yarn_noctx",
            &[
                ("llama.rope.scaling.type", Kv::Str("yarn")),
                ("llama.rope.scaling.factor", Kv::F32(8.0)),
            ],
        );
        assert!(cfg.rope_freqs.is_none());
    }

    /// `general.sampling.*` is the checkpoint's own recommendation, and
    /// only the keys the file carries become one: a file naming just
    /// `top_k` must leave temperature and top_p to the server's
    /// defaults.
    #[test]
    fn gguf_sampling_metadata_is_read_as_the_checkpoints_recommendation() {
        use crate::sampling::RecommendedSampling;
        let full = RecommendedSampling::from_gguf(&open_metadata_gguf(
            "sampling_full",
            &[
                ("general.architecture", Kv::Str("llama")),
                ("general.sampling.temp", Kv::F32(1.0)),
                ("general.sampling.top_k", Kv::U32(20)),
                ("general.sampling.top_p", Kv::F32(0.95)),
            ],
        ));
        assert_eq!(
            full,
            RecommendedSampling {
                temperature: Some(1.0),
                top_p: Some(0.95),
                top_k: Some(20),
            }
        );

        let partial = RecommendedSampling::from_gguf(&open_metadata_gguf(
            "sampling_partial",
            &[
                ("general.architecture", Kv::Str("llama")),
                ("general.sampling.top_k", Kv::U32(40)),
            ],
        ));
        assert_eq!(partial.top_k, Some(40));
        assert_eq!(partial.temperature, None);
        assert_eq!(partial.top_p, None);
    }

    /// A converter that wrote `temp = 1` stores a GGUF integer, not a
    /// float. Dropping it would serve a checkpoint that asked for
    /// temperature 1.0 at the framework's greedy default -- the
    /// repetition-loop failure the recommendation exists to prevent.
    #[test]
    fn an_integer_valued_sampling_temp_is_still_a_recommendation() {
        let recommended = crate::sampling::RecommendedSampling::from_gguf(&open_metadata_gguf(
            "sampling_int_temp",
            &[
                ("general.architecture", Kv::Str("llama")),
                ("general.sampling.temp", Kv::U32(1)),
            ],
        ));
        assert_eq!(recommended.temperature, Some(1.0));
    }

    /// The overwhelming majority of checkpoints recommend nothing, and
    /// those must keep ferrox's existing defaults exactly.
    #[test]
    fn a_gguf_without_sampling_metadata_recommends_nothing() {
        let recommended = crate::sampling::RecommendedSampling::from_gguf(&open_metadata_gguf(
            "sampling_absent",
            &[("general.architecture", Kv::Str("llama"))],
        ));
        assert!(recommended.is_empty());
    }

    #[test]
    fn model_config_from_gguf_rejects_dedicated_architectures() {
        let tmp = std::env::temp_dir().join(format!(
            "ferrox_test_dedicated_arch_{}.gguf",
            std::process::id()
        ));
        std::fs::write(&tmp, build_arch_only_gguf("deepseek4")).unwrap();
        let file = ferrox_gguf::GgufFile::open(&tmp).expect("minimal header must still parse");
        std::fs::remove_file(&tmp).ok();

        match ModelConfig::from_gguf(&file) {
            Err(LoadError::DedicatedArchitectureRequired(arch, _)) => {
                assert_eq!(arch, "deepseek4");
            }
            other => panic!(
                "expected LoadError::DedicatedArchitectureRequired for deepseek4, got {other:?}"
            ),
        }
    }

    /// The same Q5_K block bytes cross-validated against an independent
    /// Python reference in `ferrox-quant`'s own tests, reused here for
    /// the same full-path proof as the Q6_K test below.
    #[rustfmt::skip]
    const Q5_K_TEST_BLOCK: [u8; 176] = [
        0x66, 0x2a, 0x66, 0x2a, 0x01, 0x01, 0x01, 0x01, 0x4f, 0x4b, 0x10, 0x12, 0x41, 0xe2, 0xc1,
        0xb1, 0x72, 0x2f, 0x20, 0x07, 0x31, 0x0c, 0x38, 0xb3, 0x9c, 0xb8, 0xad, 0x2f, 0x9a, 0xea,
        0x17, 0xd0, 0xee, 0x93, 0x9e, 0x3e, 0x74, 0xbb, 0x28, 0x18, 0x39, 0x25, 0xb6, 0x09, 0x18,
        0x29, 0x1c, 0x1d, 0x29, 0x41, 0x40, 0x0a, 0x74, 0x7d, 0xfd, 0x21, 0xdd, 0x6d, 0x45, 0x73,
        0x0e, 0x1e, 0xc0, 0x4a, 0xfc, 0xf3, 0x8e, 0x24, 0x6b, 0x34, 0x7d, 0xbe, 0x94, 0xde, 0x59,
        0x7a, 0x35, 0x30, 0x36, 0x0a, 0xf9, 0x4a, 0x9b, 0xa2, 0x26, 0x21, 0xa2, 0xfa, 0xdf, 0x4b,
        0x29, 0x64, 0x6f, 0xbb, 0xca, 0x0f, 0x3c, 0xda, 0x20, 0xf4, 0x93, 0x86, 0xab, 0x6e, 0xb9,
        0xe5, 0xd5, 0xa0, 0x82, 0xd6, 0x41, 0xff, 0x12, 0xbc, 0x34, 0xbb, 0xab, 0xb8, 0x20, 0x2f,
        0xbb, 0x5f, 0x0c, 0x10, 0xcf, 0x49, 0xc5, 0x86, 0x5c, 0xdf, 0xff, 0x78, 0x44, 0x26, 0x3b,
        0xc2, 0x23, 0x3d, 0x2b, 0xe9, 0x00, 0x12, 0xf8, 0xea, 0xe2, 0x9e, 0x5e, 0x50, 0x20, 0x9f,
        0x9d, 0x8d, 0x7d, 0x7f, 0xcc, 0x1d, 0x0e, 0x13, 0xf8, 0xc2, 0xf1, 0x3d, 0x08, 0x2f, 0x23,
        0x13, 0xac, 0x0d, 0xa7, 0xe7, 0x20, 0xa3, 0x90, 0xb7, 0xc8, 0x28,
    ];

    fn build_single_q5_k_tensor_gguf() -> Vec<u8> {
        let mut buf = Vec::new();
        buf.write_u32::<LittleEndian>(ferrox_gguf::GGUF_MAGIC)
            .unwrap();
        buf.write_u32::<LittleEndian>(3).unwrap(); // version
        buf.write_u64::<LittleEndian>(1).unwrap(); // tensor_count
        buf.write_u64::<LittleEndian>(1).unwrap(); // kv_count

        write_kv_str(&mut buf, "general.architecture", "ferrox-q5k-test");

        write_string(&mut buf, "test.weight");
        buf.write_u32::<LittleEndian>(2).unwrap(); // n_dims
                                                   // Real GGUF ne[] order is fastest-varying-first, i.e. [cols,
                                                   // rows] -- reversed from the semantic [rows, cols] this tensor
                                                   // represents (1 row, 256 cols / 1 Q5_K block).
        buf.write_u64::<LittleEndian>(256).unwrap(); // cols (1 Q5_K block)
        buf.write_u64::<LittleEndian>(1).unwrap(); // rows
        buf.write_u32::<LittleEndian>(13).unwrap(); // dtype tag: Q5_K
        buf.write_u64::<LittleEndian>(0).unwrap(); // offset

        while buf.len() % 32 != 0 {
            buf.push(0);
        }
        buf.extend_from_slice(&Q5_K_TEST_BLOCK);
        buf
    }

    /// How far a fused dot may sit from an exact dequantized dot.
    ///
    /// Two regimes, and one fixed number cannot describe both. With
    /// `FERROX_CPU_INT_DOT` off the activation stays f32 and only
    /// rounding separates the two. With it on, the activation is
    /// quantized to int8 at `d = amax / 127`, which is the flag both
    /// binaries turn on by default and the reason the Q5_K and Q6_K
    /// cases failed against a flat `1e-2`.
    ///
    /// The bound grows with the L2 norm of the row, NOT the L1. Each
    /// element carries an independent rounding of up to `d/2`, so the
    /// dot's error is a sum of independent terms whose standard
    /// deviation is `d/sqrt(12) * ||w||_2`. Bounding by the worst case
    /// `d/2 * ||w||_1` instead assumes every rounding aligns with its
    /// weight's sign, which on this fixture gives 0.347 against a dot
    /// of 2.77: 12% of the value, loose enough that injecting a 5%
    /// error still passed. Measured here, the real error is 1.8 sigma,
    /// so four sigma keeps better than 2x headroom while still failing
    /// that 5% injection.
    fn fused_dot_tolerance(weights: &[f32], x: &[f32], exact_bound: f32) -> f32 {
        if !ferrox_core::weight_matrix::cpu_int_dot_for(
            ferrox_core::weight_matrix::IntDotShape::Matvec,
        ) {
            return exact_bound;
        }
        let amax = x.iter().fold(0.0f32, |a, v| a.max(v.abs()));
        let l2 = weights.iter().map(|w| w * w).sum::<f32>().sqrt();
        4.0 * (amax / 127.0) / 12f32.sqrt() * l2 + exact_bound
    }

    #[test]
    fn load_weight_matrix_handles_a_real_on_disk_q5_k_tensor_end_to_end() {
        let tmp = std::env::temp_dir().join(format!(
            "ferrox_test_q5k_tensor_{}.gguf",
            std::process::id()
        ));
        std::fs::write(&tmp, build_single_q5_k_tensor_gguf()).unwrap();
        let file = ferrox_gguf::GgufFile::open(&tmp).expect("real Q5_K GGUF file must parse");
        std::fs::remove_file(&tmp).ok();

        let matrix = load_weight_matrix(&file, "test.weight").expect("Q5_K tensor must load");
        assert_eq!(matrix.rows(), 1);
        assert_eq!(matrix.cols(), 256);
        match &matrix {
            WeightMatrix::Quantized { kind, data, .. } => {
                assert_eq!(*kind, QuantKind::Q5K);
                assert!(
                    data.is_mapped(),
                    "Q5_K tensors should take the zero-copy mmap path, same as Q8_0/Q4_0"
                );
            }
            _ => panic!("expected a Quantized matrix for a Q5_K tensor"),
        }

        let expected = ferrox_quant::dequant_q5_k(&Q5_K_TEST_BLOCK).unwrap();
        let x: Vec<f32> = (0..256).map(|i| ((i as f32) * 0.013).sin()).collect();
        let expected_dot: f32 = expected.iter().zip(x.iter()).map(|(a, b)| a * b).sum();

        let got = matrix.apply(&x);
        assert_eq!(got.len(), 1);
        assert!(
            (got[0] - expected_dot).abs() < fused_dot_tolerance(&expected, &x, 1e-2),
            "end-to-end loaded+applied Q5_K matrix diverged from direct dequant: got={} expected={}",
            got[0],
            expected_dot
        );
    }

    /// The same Q6_K block bytes cross-validated against an independent
    /// Python reference in `ferrox-quant`'s own tests; reused here to
    /// prove the *full*
    /// path -- real on-disk GGUF bytes, parsed by `ferrox-gguf`, read
    /// through `GgufFile::tensor_mapped_range`, dispatched by
    /// `WeightMatrix::apply` to `ferrox_quant::dot_q6_k_f32` -- produces
    /// the same result as directly dequantizing those bytes, not just
    /// that the isolated kernel is correct in unit-test isolation.
    #[rustfmt::skip]
    const Q6_K_TEST_BLOCK: [u8; 210] = [
        0xe0, 0xa5, 0x40, 0x5c, 0x8d, 0x3a, 0x0a, 0x26, 0xfb, 0x4b, 0x6e, 0x9a, 0xdf, 0x3e, 0xa3,
        0xc4, 0xf8, 0x2b, 0x1d, 0x95, 0x76, 0x7d, 0x3b, 0xcd, 0xfd, 0xef, 0xc2, 0x0b, 0x07, 0x63,
        0x29, 0xfb, 0x81, 0x57, 0xbe, 0xbe, 0x06, 0xf7, 0x3a, 0x92, 0xc4, 0x43, 0xff, 0xad, 0xac,
        0x7e, 0x0f, 0x00, 0x2a, 0x4f, 0xf0, 0xf8, 0xa9, 0xfa, 0x3c, 0x90, 0x6d, 0x73, 0x2d, 0x5a,
        0xe6, 0xc6, 0x46, 0xf2, 0x0d, 0x55, 0x4c, 0x25, 0x38, 0x71, 0x2b, 0x35, 0x38, 0x82, 0x16,
        0x37, 0x5f, 0x32, 0x61, 0x02, 0xdd, 0x2f, 0x6f, 0x7b, 0x1f, 0xb4, 0x1a, 0x1b, 0x3e, 0x4f,
        0x11, 0xa3, 0x17, 0x40, 0x5a, 0x5f, 0x76, 0xcd, 0x19, 0x27, 0x9b, 0xc7, 0xc8, 0xf7, 0xf7,
        0xee, 0xf4, 0x86, 0xd9, 0xfd, 0xa7, 0xfe, 0x9e, 0xac, 0x70, 0x53, 0x5b, 0x76, 0xfb, 0x39,
        0xf8, 0x4b, 0x98, 0xfe, 0xd0, 0x06, 0x21, 0x4c, 0x4d, 0xbe, 0x10, 0x2b, 0x06, 0x65, 0xc9,
        0x5e, 0xf9, 0x95, 0x72, 0xae, 0x99, 0xd9, 0x7e, 0x15, 0xbd, 0x5e, 0x6d, 0xe8, 0x25, 0x8a,
        0xd5, 0x99, 0xc6, 0x6b, 0x69, 0xc7, 0x84, 0xc6, 0xa4, 0xf7, 0xb9, 0x6d, 0x68, 0x45, 0x0e,
        0x65, 0x69, 0xeb, 0xe6, 0xeb, 0xe9, 0x28, 0xa6, 0xb9, 0x96, 0xf2, 0xe8, 0xa7, 0x9b, 0x6e,
        0x79, 0x8a, 0x68, 0x65, 0x59, 0x98, 0x8b, 0x44, 0x41, 0x98, 0x9a, 0x56, 0x01, 0x01, 0x01,
        0x02, 0x01, 0x01, 0x01, 0x01, 0x02, 0x01, 0x02, 0x02, 0x01, 0x01, 0x01, 0x02, 0x1f, 0x25,
    ];

    fn build_single_q6_k_tensor_gguf() -> Vec<u8> {
        let mut buf = Vec::new();
        buf.write_u32::<LittleEndian>(ferrox_gguf::GGUF_MAGIC)
            .unwrap();
        buf.write_u32::<LittleEndian>(3).unwrap(); // version
        buf.write_u64::<LittleEndian>(1).unwrap(); // tensor_count
        buf.write_u64::<LittleEndian>(1).unwrap(); // kv_count

        write_kv_str(&mut buf, "general.architecture", "ferrox-q6k-test");

        write_string(&mut buf, "test.weight");
        buf.write_u32::<LittleEndian>(2).unwrap(); // n_dims
                                                   // Real GGUF ne[] order is fastest-varying-first, i.e. [cols, rows].
        buf.write_u64::<LittleEndian>(256).unwrap(); // cols (1 Q6_K block)
        buf.write_u64::<LittleEndian>(1).unwrap(); // rows
        buf.write_u32::<LittleEndian>(14).unwrap(); // dtype tag: Q6_K
        buf.write_u64::<LittleEndian>(0).unwrap(); // offset

        while buf.len() % 32 != 0 {
            buf.push(0);
        }
        buf.extend_from_slice(&Q6_K_TEST_BLOCK);
        buf
    }

    #[test]
    fn load_weight_matrix_handles_a_real_on_disk_q6_k_tensor_end_to_end() {
        let tmp = std::env::temp_dir().join(format!(
            "ferrox_test_q6k_tensor_{}.gguf",
            std::process::id()
        ));
        std::fs::write(&tmp, build_single_q6_k_tensor_gguf()).unwrap();
        let file = ferrox_gguf::GgufFile::open(&tmp).expect("real Q6_K GGUF file must parse");
        std::fs::remove_file(&tmp).ok();

        let matrix = load_weight_matrix(&file, "test.weight").expect("Q6_K tensor must load");
        assert_eq!(matrix.rows(), 1);
        assert_eq!(matrix.cols(), 256);
        match &matrix {
            WeightMatrix::Quantized { kind, data, .. } => {
                assert_eq!(*kind, QuantKind::Q6K);
                assert!(
                    data.is_mapped(),
                    "Q6_K tensors should take the zero-copy mmap path, same as Q8_0/Q4_0"
                );
            }
            _ => panic!("expected a Quantized matrix for a Q6_K tensor"),
        }

        let expected = ferrox_quant::dequant_q6_k(&Q6_K_TEST_BLOCK).unwrap();
        let x: Vec<f32> = (0..256).map(|i| ((i as f32) * 0.013).sin()).collect();
        let expected_dot: f32 = expected.iter().zip(x.iter()).map(|(a, b)| a * b).sum();

        let got = matrix.apply(&x);
        assert_eq!(got.len(), 1);
        assert!(
            (got[0] - expected_dot).abs() < fused_dot_tolerance(&expected, &x, 1e-2),
            "end-to-end loaded+applied Q6_K matrix diverged from direct dequant: got={} expected={}",
            got[0],
            expected_dot
        );
    }

    fn build_single_bf16_tensor_gguf(rows: u64, cols: u64, values: &[f32]) -> Vec<u8> {
        let mut buf = Vec::new();
        buf.write_u32::<LittleEndian>(ferrox_gguf::GGUF_MAGIC)
            .unwrap();
        buf.write_u32::<LittleEndian>(3).unwrap(); // version
        buf.write_u64::<LittleEndian>(1).unwrap(); // tensor_count
        buf.write_u64::<LittleEndian>(1).unwrap(); // kv_count

        write_kv_str(&mut buf, "general.architecture", "ferrox-bf16-test");

        write_string(&mut buf, "test.weight");
        buf.write_u32::<LittleEndian>(2).unwrap(); // n_dims
                                                   // Real GGUF ne[] order is fastest-varying-first, i.e. [cols, rows].
        buf.write_u64::<LittleEndian>(cols).unwrap();
        buf.write_u64::<LittleEndian>(rows).unwrap();
        buf.write_u32::<LittleEndian>(30).unwrap(); // dtype tag: BF16
        buf.write_u64::<LittleEndian>(0).unwrap(); // offset

        while buf.len() % 32 != 0 {
            buf.push(0);
        }
        for &v in values {
            // Real bf16 truncation (round-toward-zero, matching a real
            // writer closely enough for round-trip test purposes): top
            // 16 bits of the f32 bit pattern.
            let bf16_bits = (v.to_bits() >> 16) as u16;
            buf.extend_from_slice(&bf16_bits.to_le_bytes());
        }
        buf
    }

    #[test]
    fn load_weight_matrix_handles_a_real_on_disk_bf16_tensor_end_to_end() {
        // Values with zero low-mantissa bits, so f32->bf16 truncation
        // is lossless and this is an exact-equality check.
        let values: Vec<f32> = vec![1.0, -2.5, 0.0, 4.0, -8.0, 16.0];
        let tmp = std::env::temp_dir().join(format!(
            "ferrox_test_bf16_tensor_{}.gguf",
            std::process::id()
        ));
        std::fs::write(&tmp, build_single_bf16_tensor_gguf(2, 3, &values)).unwrap();
        let file = ferrox_gguf::GgufFile::open(&tmp).expect("real BF16 GGUF file must parse");
        std::fs::remove_file(&tmp).ok();

        let matrix = load_weight_matrix(&file, "test.weight").expect("BF16 tensor must load");
        assert_eq!(matrix.rows(), 2);
        assert_eq!(matrix.cols(), 3);
        match &matrix {
            WeightMatrix::F32(tensor) => {
                assert_eq!(tensor.data, values, "BF16 must widen to f32 exactly");
            }
            _ => panic!("expected an F32 matrix for a BF16 tensor (no fused dot kernel for it)"),
        }
    }

    fn build_single_f16_tensor_gguf(rows: u64, cols: u64, values: &[f32]) -> Vec<u8> {
        let mut buf = Vec::new();
        buf.write_u32::<LittleEndian>(ferrox_gguf::GGUF_MAGIC)
            .unwrap();
        buf.write_u32::<LittleEndian>(3).unwrap(); // version
        buf.write_u64::<LittleEndian>(1).unwrap(); // tensor_count
        buf.write_u64::<LittleEndian>(1).unwrap(); // kv_count

        write_kv_str(&mut buf, "general.architecture", "ferrox-f16-test");

        write_string(&mut buf, "test.weight");
        buf.write_u32::<LittleEndian>(2).unwrap(); // n_dims
        buf.write_u64::<LittleEndian>(cols).unwrap();
        buf.write_u64::<LittleEndian>(rows).unwrap();
        buf.write_u32::<LittleEndian>(1).unwrap(); // dtype tag: F16
        buf.write_u64::<LittleEndian>(0).unwrap(); // offset

        while buf.len() % 32 != 0 {
            buf.push(0);
        }
        for &v in values {
            buf.extend_from_slice(&half::f16::from_f32(v).to_le_bytes());
        }
        buf
    }

    /// `GgmlType::F16` was parsed and sized but had no dequant arm in any
    /// of the seven loaders, so every `*-f16.gguf` was a hard
    /// `UnsupportedDtype`. Values are exactly representable in f16, so
    /// this is an exact-equality check.
    #[test]
    fn load_weight_matrix_handles_a_real_on_disk_f16_tensor_end_to_end() {
        let values: Vec<f32> = vec![1.0, -2.5, 0.0, 4.0, -8.0, 16.0];
        let tmp = std::env::temp_dir().join(format!(
            "ferrox_test_f16_tensor_{}.gguf",
            std::process::id()
        ));
        std::fs::write(&tmp, build_single_f16_tensor_gguf(2, 3, &values)).unwrap();
        let file = ferrox_gguf::GgufFile::open(&tmp).expect("real F16 GGUF file must parse");
        std::fs::remove_file(&tmp).ok();

        let matrix = load_weight_matrix(&file, "test.weight").expect("F16 tensor must load");
        assert_eq!(matrix.rows(), 2);
        assert_eq!(matrix.cols(), 3);
        match &matrix {
            WeightMatrix::F32(tensor) => {
                assert_eq!(tensor.data, values, "F16 must widen to f32 exactly");
            }
            _ => panic!("expected an F32 matrix for an F16 tensor (no fused dot kernel for it)"),
        }

        // The same tensor read as a plain vector (norm weights, biases and
        // the router all take this path, not `load_weight_matrix`).
        let tmp =
            std::env::temp_dir().join(format!("ferrox_test_f16_vec_{}.gguf", std::process::id()));
        std::fs::write(&tmp, build_single_f16_tensor_gguf(2, 3, &values)).unwrap();
        let file = ferrox_gguf::GgufFile::open(&tmp).expect("real F16 GGUF file must parse");
        std::fs::remove_file(&tmp).ok();
        assert_eq!(load_f32_vec(&file, "test.weight").unwrap(), values);
    }

    fn build_single_q5_1_tensor_gguf() -> Vec<u8> {
        let mut buf = Vec::new();
        buf.write_u32::<LittleEndian>(ferrox_gguf::GGUF_MAGIC)
            .unwrap();
        buf.write_u32::<LittleEndian>(3).unwrap(); // version
        buf.write_u64::<LittleEndian>(1).unwrap(); // tensor_count
        buf.write_u64::<LittleEndian>(1).unwrap(); // kv_count

        write_kv_str(&mut buf, "general.architecture", "ferrox-q5-1-test");

        write_string(&mut buf, "test.weight");
        buf.write_u32::<LittleEndian>(2).unwrap(); // n_dims
                                                   // Real GGUF ne[] order is fastest-varying-first, i.e. [cols, rows].
        buf.write_u64::<LittleEndian>(32).unwrap(); // cols (1 Q5_1 block)
        buf.write_u64::<LittleEndian>(1).unwrap(); // rows
        buf.write_u32::<LittleEndian>(7).unwrap(); // dtype tag: Q5_1
        buf.write_u64::<LittleEndian>(0).unwrap(); // offset

        while buf.len() % 32 != 0 {
            buf.push(0);
        }
        // d=0.25 (f16 0x3400), m=1.5 (f16 0x3E00) -- both exact in f16,
        // hand-verified bit patterns to avoid pulling in the `half`
        // crate just for two test constants. qh varied, qs a real
        // (non-degenerate) pattern.
        buf.extend_from_slice(&0x3400u16.to_le_bytes());
        buf.extend_from_slice(&0x3E00u16.to_le_bytes());
        buf.extend_from_slice(&[0x9au8, 0x3c, 0xf0, 0x0f]);
        buf.extend_from_slice(&(0..16u8).map(|i| i | ((15 - i) << 4)).collect::<Vec<u8>>());
        buf
    }

    #[test]
    fn load_weight_matrix_handles_a_real_on_disk_q5_1_tensor_end_to_end() {
        let tmp = std::env::temp_dir().join(format!(
            "ferrox_test_q5_1_tensor_{}.gguf",
            std::process::id()
        ));
        std::fs::write(&tmp, build_single_q5_1_tensor_gguf()).unwrap();
        let file = ferrox_gguf::GgufFile::open(&tmp).expect("real Q5_1 GGUF file must parse");
        std::fs::remove_file(&tmp).ok();

        let matrix = load_weight_matrix(&file, "test.weight").expect("Q5_1 tensor must load");
        assert_eq!(matrix.rows(), 1);
        assert_eq!(matrix.cols(), 32);
        let raw = file.tensor_bytes("test.weight").unwrap();
        let expected = ferrox_quant::dequant_q5_1(raw).unwrap();
        match &matrix {
            WeightMatrix::Quantized { kind, data, .. } => {
                assert_eq!(*kind, QuantKind::Q5_1);
                assert!(data.is_mapped());
            }
            _ => panic!("expected a Quantized matrix for a Q5_1 tensor"),
        }

        let x: Vec<f32> = (0..32).map(|i| ((i as f32) * 0.017).cos()).collect();
        let expected_dot: f32 = expected.iter().zip(x.iter()).map(|(a, b)| a * b).sum();
        let got = matrix.apply(&x);
        assert_eq!(got.len(), 1);
        assert!(
            (got[0] - expected_dot).abs() < 1e-2,
            "end-to-end loaded+applied Q5_1 matrix diverged from direct dequant: got={} expected={}",
            got[0],
            expected_dot
        );
    }

    // Same bytes as ferrox-quant's own Q3_K_TEST_BLOCK (Python-cross-
    // validated there); duplicated here to build a real on-disk GGUF
    // file, matching this file's existing per-format test convention
    // (see Q6_K_TEST_BLOCK above).
    const Q3_K_TEST_BLOCK: [u8; 110] = [
        0x56, 0xf2, 0xb4, 0x2b, 0xd5, 0x6f, 0x51, 0x71, 0x3c, 0x0a, 0xb9, 0x1d, 0xd0, 0xb9, 0x3b,
        0xb3, 0x0f, 0xff, 0x8c, 0xb2, 0x83, 0x3a, 0x3d, 0x24, 0xb1, 0x12, 0x56, 0xe3, 0x23, 0x54,
        0xf2, 0xfa, 0x7f, 0xdf, 0x31, 0xe1, 0x18, 0x26, 0x6e, 0xcd, 0x5b, 0x38, 0xee, 0xbd, 0x9f,
        0x8c, 0x57, 0x47, 0x0b, 0x11, 0xcb, 0xfb, 0xb4, 0x83, 0xa0, 0x4e, 0x0b, 0xd4, 0xa7, 0x85,
        0xe0, 0x60, 0xf3, 0xb3, 0xe3, 0x95, 0x43, 0xc6, 0x05, 0x05, 0x77, 0x53, 0xed, 0x23, 0xcc,
        0x6a, 0x0e, 0x89, 0xa1, 0x79, 0x85, 0xf6, 0x6e, 0x5a, 0x23, 0x63, 0xbe, 0x53, 0xfa, 0xa2,
        0x2b, 0xe9, 0xcd, 0xce, 0xf8, 0x3d, 0x6f, 0xd0, 0x42, 0x6e, 0x3b, 0x7f, 0x23, 0x26, 0xd3,
        0xb9, 0x18, 0xbf, 0xa4, 0x34,
    ];

    fn build_single_q3_k_tensor_gguf() -> Vec<u8> {
        let mut buf = Vec::new();
        buf.write_u32::<LittleEndian>(ferrox_gguf::GGUF_MAGIC)
            .unwrap();
        buf.write_u32::<LittleEndian>(3).unwrap(); // version
        buf.write_u64::<LittleEndian>(1).unwrap(); // tensor_count
        buf.write_u64::<LittleEndian>(1).unwrap(); // kv_count

        write_kv_str(&mut buf, "general.architecture", "ferrox-q3k-test");

        write_string(&mut buf, "test.weight");
        buf.write_u32::<LittleEndian>(2).unwrap(); // n_dims
                                                   // Real GGUF ne[] order is fastest-varying-first, i.e. [cols, rows].
        buf.write_u64::<LittleEndian>(256).unwrap(); // cols (1 Q3_K block)
        buf.write_u64::<LittleEndian>(1).unwrap(); // rows
        buf.write_u32::<LittleEndian>(11).unwrap(); // dtype tag: Q3_K
        buf.write_u64::<LittleEndian>(0).unwrap(); // offset

        while buf.len() % 32 != 0 {
            buf.push(0);
        }
        buf.extend_from_slice(&Q3_K_TEST_BLOCK);
        buf
    }

    #[test]
    fn load_weight_matrix_handles_a_real_on_disk_q3_k_tensor_end_to_end() {
        let tmp = std::env::temp_dir().join(format!(
            "ferrox_test_q3k_tensor_{}.gguf",
            std::process::id()
        ));
        std::fs::write(&tmp, build_single_q3_k_tensor_gguf()).unwrap();
        let file = ferrox_gguf::GgufFile::open(&tmp).expect("real Q3_K GGUF file must parse");
        std::fs::remove_file(&tmp).ok();

        let matrix = load_weight_matrix(&file, "test.weight").expect("Q3_K tensor must load");
        assert_eq!(matrix.rows(), 1);
        assert_eq!(matrix.cols(), 256);
        match &matrix {
            WeightMatrix::Quantized { kind, data, .. } => {
                assert_eq!(*kind, QuantKind::Q3K);
                assert!(data.is_mapped());
            }
            _ => panic!("expected a Quantized matrix for a Q3_K tensor"),
        }

        let expected = ferrox_quant::dequant_q3_k(&Q3_K_TEST_BLOCK).unwrap();
        let x: Vec<f32> = (0..256).map(|i| ((i as f32) * 0.013).sin()).collect();
        let expected_dot: f32 = expected.iter().zip(x.iter()).map(|(a, b)| a * b).sum();

        let got = matrix.apply(&x);
        assert_eq!(got.len(), 1);
        assert!(
            (got[0] - expected_dot).abs() < fused_dot_tolerance(&expected, &x, 1e-1),
            "end-to-end loaded+applied Q3_K matrix diverged from direct dequant: got={} expected={}",
            got[0],
            expected_dot
        );
    }

    // Same bytes as ferrox-quant's own IQ4_XS_TEST_BLOCK (Python-cross-
    // validated there); duplicated here to build a real on-disk GGUF
    // file, matching this file's existing per-format test convention.
    const IQ4_XS_TEST_BLOCK: [u8; 136] = [
        0x5c, 0x33, 0xb4, 0x39, 0xd1, 0x64, 0x97, 0x82, 0xcb, 0xbd, 0x88, 0x95, 0xf3, 0x60, 0x2a,
        0xb5, 0xe7, 0x24, 0xd3, 0xee, 0xfe, 0x71, 0x13, 0xbe, 0x70, 0x84, 0x48, 0x79, 0x7b, 0x3e,
        0xf0, 0x55, 0xdc, 0xb2, 0xb2, 0xde, 0x32, 0xa1, 0x5b, 0x02, 0x01, 0xdc, 0x2a, 0xbb, 0xf7,
        0x0b, 0x8a, 0x88, 0xdd, 0x0b, 0x02, 0x7e, 0x5e, 0x76, 0x87, 0x30, 0x1e, 0x1c, 0xcf, 0x48,
        0xd7, 0x61, 0xf3, 0x51, 0x52, 0x17, 0x98, 0x0a, 0x87, 0xcf, 0x02, 0x91, 0xc8, 0xee, 0xc0,
        0x91, 0x69, 0x2a, 0x4f, 0x64, 0x68, 0xa7, 0xb2, 0xe6, 0x98, 0x21, 0x81, 0x75, 0x53, 0x2a,
        0x8d, 0x12, 0xae, 0xe0, 0xea, 0x0c, 0x75, 0xff, 0x22, 0x5e, 0x25, 0x19, 0xda, 0x2e, 0x51,
        0x4e, 0x81, 0xdc, 0x0e, 0x78, 0x86, 0xd7, 0x58, 0xb5, 0xb7, 0xf6, 0x45, 0xa9, 0x0a, 0x83,
        0xfd, 0x2a, 0x12, 0x7d, 0xf0, 0x12, 0x97, 0xe2, 0xfe, 0xf4, 0xd0, 0xa2, 0x11, 0x14, 0x78,
        0xdb,
    ];

    fn build_single_iq4_xs_tensor_gguf() -> Vec<u8> {
        let mut buf = Vec::new();
        buf.write_u32::<LittleEndian>(ferrox_gguf::GGUF_MAGIC)
            .unwrap();
        buf.write_u32::<LittleEndian>(3).unwrap(); // version
        buf.write_u64::<LittleEndian>(1).unwrap(); // tensor_count
        buf.write_u64::<LittleEndian>(1).unwrap(); // kv_count

        write_kv_str(&mut buf, "general.architecture", "ferrox-iq4xs-test");

        write_string(&mut buf, "test.weight");
        buf.write_u32::<LittleEndian>(2).unwrap(); // n_dims
                                                   // Real GGUF ne[] order is fastest-varying-first, i.e. [cols, rows].
        buf.write_u64::<LittleEndian>(256).unwrap(); // cols (1 IQ4_XS block)
        buf.write_u64::<LittleEndian>(1).unwrap(); // rows
        buf.write_u32::<LittleEndian>(23).unwrap(); // dtype tag: IQ4_XS
        buf.write_u64::<LittleEndian>(0).unwrap(); // offset

        while buf.len() % 32 != 0 {
            buf.push(0);
        }
        buf.extend_from_slice(&IQ4_XS_TEST_BLOCK);
        buf
    }

    #[test]
    fn load_weight_matrix_handles_a_real_on_disk_iq4_xs_tensor_end_to_end() {
        let tmp = std::env::temp_dir().join(format!(
            "ferrox_test_iq4xs_tensor_{}.gguf",
            std::process::id()
        ));
        std::fs::write(&tmp, build_single_iq4_xs_tensor_gguf()).unwrap();
        let file = ferrox_gguf::GgufFile::open(&tmp).expect("real IQ4_XS GGUF file must parse");
        std::fs::remove_file(&tmp).ok();

        let matrix = load_weight_matrix(&file, "test.weight").expect("IQ4_XS tensor must load");
        assert_eq!(matrix.rows(), 1);
        assert_eq!(matrix.cols(), 256);
        match &matrix {
            WeightMatrix::Quantized { kind, data, .. } => {
                assert_eq!(*kind, QuantKind::IQ4XS);
                assert!(data.is_mapped());
            }
            _ => panic!("expected a Quantized matrix for an IQ4_XS tensor"),
        }

        let expected = ferrox_quant::dequant_iq4_xs(&IQ4_XS_TEST_BLOCK).unwrap();
        let x: Vec<f32> = (0..256).map(|i| ((i as f32) * 0.013).sin()).collect();
        let expected_dot: f32 = expected.iter().zip(x.iter()).map(|(a, b)| a * b).sum();

        let got = matrix.apply(&x);
        assert_eq!(got.len(), 1);
        assert!(
            (got[0] - expected_dot).abs() < 1e-1,
            "end-to-end loaded+applied IQ4_XS matrix diverged from direct dequant: got={} expected={}",
            got[0],
            expected_dot
        );
    }

    // Same bytes as ferrox-quant's own IQ low-bit test blocks
    // (Python-cross-validated there against the real compiled ggml
    // implementation), duplicated as literals for the same reason as
    // IQ4_XS_TEST_BLOCK above.
    const IQ1_S_TEST_BLOCK: [u8; 50] = [
        0x0a, 0x2f, 0xfa, 0x06, 0x1e, 0x37, 0x6f, 0xe3, 0x62, 0xd0, 0xb6, 0xa4, 0x25, 0xae, 0x76,
        0x14, 0x72, 0x5b, 0xfa, 0x05, 0xd1, 0xf1, 0x2a, 0x4c, 0xad, 0x29, 0xae, 0xf4, 0xcf, 0x0c,
        0x96, 0x51, 0x58, 0x03, 0x6d, 0xd3, 0x10, 0x92, 0x70, 0xff, 0x61, 0x58, 0xc8, 0x30, 0x25,
        0x64, 0x49, 0x85, 0xc0, 0x24,
    ];
    const IQ2_XXS_TEST_BLOCK: [u8; 66] = [
        0x29, 0x30, 0xd9, 0x33, 0x95, 0x4c, 0x08, 0x1e, 0xad, 0x79, 0x49, 0xf2, 0x8d, 0x5f, 0x93,
        0xea, 0x78, 0x18, 0x98, 0xb9, 0x94, 0x14, 0xad, 0xce, 0xca, 0x1d, 0xab, 0x81, 0x53, 0x4a,
        0x68, 0xd0, 0x59, 0x96, 0x36, 0x5d, 0xbe, 0x20, 0xc4, 0xff, 0xe4, 0x2c, 0xcd, 0x2f, 0x4f,
        0x4f, 0x67, 0x53, 0xc6, 0xd5, 0xa2, 0xfb, 0xc7, 0xf3, 0xe2, 0x6b, 0xf1, 0x99, 0x23, 0x1e,
        0x2d, 0x5e, 0x8c, 0x78, 0xc2, 0x31,
    ];
    const IQ3_XXS_TEST_BLOCK: [u8; 98] = [
        0x71, 0x31, 0x16, 0x0a, 0x79, 0x04, 0x5d, 0x87, 0xae, 0x2a, 0x4a, 0x43, 0xfd, 0x02, 0xba,
        0x6c, 0x10, 0x42, 0x80, 0xe5, 0x1d, 0x08, 0x22, 0xcb, 0x21, 0x54, 0xf9, 0xaa, 0x8e, 0xc2,
        0xf2, 0x34, 0x66, 0x1e, 0x2a, 0xef, 0x19, 0xae, 0x48, 0x47, 0x29, 0xa0, 0x72, 0xd1, 0x31,
        0xc0, 0x65, 0x49, 0xde, 0x79, 0x32, 0xe6, 0x4d, 0xb6, 0x55, 0x3f, 0x4d, 0xf1, 0x18, 0xbb,
        0x18, 0x59, 0x4c, 0x31, 0xa3, 0xb2, 0x34, 0xdd, 0xf6, 0x4a, 0x91, 0x51, 0x3f, 0x3e, 0x40,
        0x69, 0xad, 0xbf, 0x1a, 0xd0, 0x05, 0xfb, 0xbe, 0x8b, 0x0b, 0xdd, 0xdf, 0x7d, 0x94, 0x74,
        0x92, 0x3e, 0xff, 0x04, 0x2a, 0xc4, 0xea, 0xc9,
    ];

    #[rustfmt::skip]
    const MXFP4_GGUF_TEST_BLOCKS: [u8; 68] = [0x79, 0xb4, 0x8d, 0xe2, 0x62, 0x5d, 0xbb, 0x9d, 0x54, 0xe6, 0xdb, 0x94, 0x59, 0x7d, 0x28, 0xf9, 0x79, 0x7a, 0xfc, 0xc1, 0xfa, 0x1e, 0x53, 0x5b, 0x0e, 0xc2, 0x5a, 0x2f, 0x0c, 0x82, 0x4d, 0xcb, 0x11, 0x28, 0x7b, 0x7c, 0xb6, 0x45, 0xe0, 0xb0, 0x52, 0x40, 0x51, 0xec, 0x30, 0x1a, 0xd2, 0x17, 0xf3, 0xbb, 0xfc, 0x7c, 0x8f, 0xf0, 0x67, 0x83, 0x88, 0x9d, 0x79, 0xdb, 0xf4, 0x45, 0x29, 0x78, 0xe6, 0xf4, 0x99, 0xea];

    /// A live ggml type this build has no kernel for must be REFUSED BY
    /// NAME at execution, having been sized correctly at parse.
    ///
    /// Before `TQ2_0` was recognized, tag 35 was `Other(35)`, which had
    /// no block layout: the tensor's size was unknown, so `tensor_bytes`
    /// could not even hand back the row, and the error named a number.
    /// Now the file parses, the tensor measures 66 bytes per 256
    /// elements, and the stop happens where it belongs -- at the point
    /// something wants to multiply by it -- naming `TQ2_0`.
    #[test]
    fn a_recognized_but_unimplemented_ggml_type_refuses_by_name_after_sizing_correctly() {
        // 256 elements of TQ2_0 = one 66-byte block.
        let block = pseudo_iq_block(66, 0x0720_5eed);
        let tmp =
            std::env::temp_dir().join(format!("ferrox_test_tq2_0_{}.gguf", std::process::id()));
        std::fs::write(
            &tmp,
            build_single_iq_lowbit_tensor_gguf("tq2test", 35, 256, &block),
        )
        .unwrap();
        let file = ferrox_gguf::GgufFile::open(&tmp).expect("a TQ2_0 file must still parse");
        std::fs::remove_file(&tmp).ok();

        // Sized, not zero: the size estimate is right even though the
        // kernel is missing.
        let info = file.find_tensor("test.weight").expect("tensor present");
        assert_eq!(info.dtype, GgmlType::TQ2_0);
        assert_eq!(info.byte_len(), Some(66));
        assert_eq!(
            file.tensor_bytes("test.weight").map(<[u8]>::len).ok(),
            Some(66)
        );

        match load_weight_matrix(&file, "test.weight") {
            Err(LoadError::UnsupportedDtype(name, GgmlType::TQ2_0)) => {
                assert_eq!(name, "test.weight");
            }
            Err(other) => panic!("TQ2_0 must be refused by name, got {other:?}"),
            Ok(_) => panic!("TQ2_0 must be refused, not loaded as some other kind"),
        }
    }

    /// An MXFP4 norm/bias must widen, not be refused.
    ///
    /// `load_weight_matrix` accepts MXFP4 as a 2-D weight and
    /// `load_moe_expert_matrices` accepts it as an expert tensor, and
    /// `WeightMatrix::dequant` calls `dequant_mxfp4_gguf` on both. One
    /// missing arm in `widen_plain_float` made the *1-D* tensors of the
    /// exact same dtype a hard `UnsupportedDtype` -- the split that
    /// turns a supported format into a load failure on the one
    /// checkpoint that uses it.
    #[test]
    fn an_mxfp4_one_dimensional_tensor_widens_instead_of_being_refused() {
        let expected = ferrox_quant::dequant_mxfp4_gguf(&MXFP4_GGUF_TEST_BLOCKS)
            .expect("the fixture blocks must dequantize");
        let cols = expected.len();
        let tmp = std::env::temp_dir().join(format!(
            "ferrox_test_mxfp4_norm_{}.gguf",
            std::process::id()
        ));
        std::fs::write(
            &tmp,
            build_single_iq_lowbit_tensor_gguf(
                "mxfp4norm",
                39,
                cols as u64,
                &MXFP4_GGUF_TEST_BLOCKS,
            ),
        )
        .unwrap();
        let file = ferrox_gguf::GgufFile::open(&tmp).expect("file must parse");
        std::fs::remove_file(&tmp).ok();

        let got = load_f32_vec(&file, "test.weight")
            .expect("an MXFP4 norm must load, not report an unsupported dtype");
        assert_eq!(got, expected);

        // Same arm, reached directly: `widen_plain_float` is the shared
        // helper the six architecture loaders call, so its table is the
        // one that has to know MXFP4.
        let direct = widen_plain_float(GgmlType::MXFP4, &MXFP4_GGUF_TEST_BLOCKS, "test.weight")
            .expect("widen_plain_float must widen MXFP4");
        assert_eq!(direct, expected);

        // And the refusal still works for a dtype that genuinely has no
        // widening path, so this test cannot pass by making everything
        // succeed.
        match widen_plain_float(GgmlType::TQ2_0, &MXFP4_GGUF_TEST_BLOCKS, "test.weight") {
            Err(LoadError::UnsupportedDtype(name, GgmlType::TQ2_0)) => {
                assert_eq!(name, "test.weight");
            }
            other => panic!("TQ2_0 must be refused by name, got {other:?}"),
        }
    }

    fn build_single_iq_lowbit_tensor_gguf(
        arch: &str,
        tag: u32,
        cols: u64,
        block: &[u8],
    ) -> Vec<u8> {
        let mut buf = Vec::new();
        buf.write_u32::<LittleEndian>(ferrox_gguf::GGUF_MAGIC)
            .unwrap();
        buf.write_u32::<LittleEndian>(3).unwrap(); // version
        buf.write_u64::<LittleEndian>(1).unwrap(); // tensor_count
        buf.write_u64::<LittleEndian>(1).unwrap(); // kv_count
        write_kv_str(&mut buf, "general.architecture", arch);
        write_string(&mut buf, "test.weight");
        buf.write_u32::<LittleEndian>(2).unwrap(); // n_dims
        buf.write_u64::<LittleEndian>(cols).unwrap();
        buf.write_u64::<LittleEndian>(1).unwrap(); // rows
        buf.write_u32::<LittleEndian>(tag).unwrap();
        buf.write_u64::<LittleEndian>(0).unwrap(); // offset
        while buf.len() % 32 != 0 {
            buf.push(0);
        }
        buf.extend_from_slice(block);
        buf
    }

    /// A structurally valid block of `len` bytes for any of the
    /// codebook-grid formats: every bit pattern is a legal code in all
    /// of them (the grid indices are bounded by their own bit widths),
    /// so a deterministic byte fill is a real block, not a fixture that
    /// happens to avoid the interesting paths. Only the f16 scale needs
    /// pinning, and only so the comparison below can't be NaN-vs-NaN.
    fn pseudo_iq_block(len: usize, seed: u32) -> Vec<u8> {
        let mut s = seed;
        let mut out = Vec::with_capacity(len);
        for _ in 0..len {
            s ^= s << 13;
            s ^= s >> 17;
            s ^= s << 5;
            out.push((s >> 24) as u8);
        }
        out
    }

    /// End-to-end load+apply for the codebook-grid low-bit formats the
    /// published Dynamic GGUFs are built from: a real on-disk tensor of
    /// each type must load zero-copy as the right `QuantKind` and
    /// produce the same matvec result as dequantizing the block
    /// directly. That is the property this test exists for -- the
    /// *values* are pinned against real ggml in `ferrox-quant`; what
    /// can only break here is the tag -> kind -> block-stride chain,
    /// and a wrong stride silently reads the neighbouring row.
    /// Dtype tags (19/29/16/17/22/18/21/39) verified against ggml.h's
    /// enum ggml_type.
    #[test]
    fn load_weight_matrix_handles_real_on_disk_iq_lowbit_tensors_end_to_end() {
        type DequantFn = fn(&[u8]) -> Result<Vec<f32>, ferrox_quant::QuantError>;
        // IQ1_M carries no f16 scale field; its scale is reassembled
        // from the four scale words' top nibbles, and the top nibble of
        // the last one supplies the f16 sign + high exponent bits.
        // Pinning it to 0x2 keeps the exponent out of the all-ones
        // NaN/Inf pattern whatever the rest of the fill does. The other
        // three do carry a leading f16 `d`, pinned for the same reason.
        let mut iq1m = pseudo_iq_block(ferrox_quant::IQ1_M_BLOCK_BYTES, 0x2907_31A0);
        iq1m[55] = (iq1m[55] & 0x0F) | 0x20;
        let mut iq2xs = pseudo_iq_block(ferrox_quant::IQ2_XS_BLOCK_BYTES, 0x2107_31A1);
        let mut iq2s = pseudo_iq_block(ferrox_quant::IQ2_S_BLOCK_BYTES, 0x2207_31A2);
        let mut iq3s = pseudo_iq_block(ferrox_quant::IQ3_S_BLOCK_BYTES, 0x2307_31A3);
        for blk in [&mut iq2xs, &mut iq2s, &mut iq3s] {
            blk[0..2].copy_from_slice(&half::f16::from_f32(0.115).to_le_bytes());
        }
        let cases: [(&str, u32, &[u8], QuantKind, DequantFn); 8] = [
            (
                "iq1s",
                19,
                &IQ1_S_TEST_BLOCK,
                QuantKind::IQ1S,
                ferrox_quant::dequant_iq1_s,
            ),
            (
                "iq1m",
                29,
                &iq1m,
                QuantKind::IQ1M,
                ferrox_quant::dequant_iq1_m,
            ),
            (
                "iq2xxs",
                16,
                &IQ2_XXS_TEST_BLOCK,
                QuantKind::IQ2XXS,
                ferrox_quant::dequant_iq2_xxs,
            ),
            (
                "iq2xs",
                17,
                &iq2xs,
                QuantKind::IQ2XS,
                ferrox_quant::dequant_iq2_xs,
            ),
            (
                "iq2s",
                22,
                &iq2s,
                QuantKind::IQ2S,
                ferrox_quant::dequant_iq2_s,
            ),
            (
                "iq3xxs",
                18,
                &IQ3_XXS_TEST_BLOCK,
                QuantKind::IQ3XXS,
                ferrox_quant::dequant_iq3_xxs,
            ),
            (
                "iq3s",
                21,
                &iq3s,
                QuantKind::IQ3S,
                ferrox_quant::dequant_iq3_s,
            ),
            (
                "mxfp4_gguf",
                39,
                &MXFP4_GGUF_TEST_BLOCKS,
                QuantKind::Mxfp4Gguf,
                ferrox_quant::dequant_mxfp4_gguf,
            ),
        ];
        for (name, tag, block, kind, dequant) in cases {
            let expected = dequant(block).unwrap();
            let cols = expected.len();
            let tmp = std::env::temp_dir().join(format!("ferrox_test_{name}_tensor.gguf"));
            std::fs::write(
                &tmp,
                build_single_iq_lowbit_tensor_gguf(name, tag, cols as u64, block),
            )
            .unwrap();
            let file = ferrox_gguf::GgufFile::open(&tmp).expect("file must parse");
            std::fs::remove_file(&tmp).ok();

            let matrix =
                load_weight_matrix(&file, "test.weight").expect("low-bit tensor must load");
            assert_eq!((matrix.rows(), matrix.cols()), (1, cols), "{name}");
            match &matrix {
                WeightMatrix::Quantized { kind: k, data, .. } => {
                    assert_eq!(*k, kind, "{name}");
                    assert!(data.is_mapped(), "{name} must load zero-copy");
                }
                _ => panic!("expected a Quantized matrix for {name}"),
            }

            let x: Vec<f32> = (0..cols).map(|i| ((i as f32) * 0.013).sin()).collect();
            let expected_dot: f32 = expected.iter().zip(x.iter()).map(|(a, b)| a * b).sum();
            let got = matrix.apply(&x);
            assert!(
                (got[0] - expected_dot).abs() < 1e-1,
                "{name}: loaded+applied diverged from direct dequant: got={} expected={}",
                got[0],
                expected_dot
            );
        }
    }

    #[test]
    fn qwen2moe_disables_topk_renorm() {
        assert!(
            NO_TOPK_RENORMALIZE_ARCHITECTURES.contains(&"qwen2moe"),
            "qwen2moe must have norm_topk_prob=false (llama.cpp build_moe_ffn norm_w=false)"
        );
    }

    /// An architecture that uses no RoPE must not reach the generic
    /// decoder, which rotates unconditionally.
    ///
    /// All five of these were admitted as `GenericGqa { rope: Neox }`.
    /// Nothing downstream could have caught it: `bloom` and `refact`
    /// hardcode their ALiBi slope in `load_arch_hparams` with no GGUF
    /// key, so the metadata gates above see nothing, and `mpt` carries
    /// no tensor the generic loader fails to consume, so
    /// `assert_every_tensor_consumed` sees nothing either. It would have
    /// loaded, run at full speed, and answered from rotated positions.
    #[test]
    fn an_architecture_with_no_rope_is_refused_by_name() {
        for arch in ["gpt2", "mpt", "refact", "bloom", "jais"] {
            let file = open_metadata_gguf(
                &format!("norope_{arch}"),
                &[("general.architecture", Kv::Str(arch))],
            );
            match ModelConfig::from_gguf(&file) {
                Err(LoadError::DedicatedArchitectureRequired(got, reason)) => {
                    assert_eq!(got, arch);
                    assert!(
                        reason.contains("ALiBi") || reason.contains("position embeddings"),
                        "{arch}: the refusal must name what is missing, got {reason:?}"
                    );
                }
                other => panic!("{arch} must be refused, got {other:?}"),
            }
        }
    }

    /// A per-layer sliding-window pattern is refused, and a scalar one
    /// still loads.
    ///
    /// Both halves matter. `capability::unsupported_feature_keys` used
    /// to refuse the key outright with the reason "not implemented in
    /// the generic decoder", which was false -- the alternating pattern
    /// is `ModelConfig::layer_sliding_window`, both phases, and
    /// `gpt-oss` has run on it since it was audited. What that gate
    /// really did was make the loader's own read of the key dead for
    /// every non-Gemma architecture. The case ferrox genuinely cannot
    /// express is the ARRAY, which `metadata_u64_any` reads as `None`
    /// and which therefore used to fall through to
    /// `default_swa_layout` -- substituting the architecture's
    /// hardcoded layout for the file's, silently.
    #[test]
    fn an_array_valued_sliding_window_pattern_is_refused_rather_than_substituted() {
        // llama.cpp reads this key with `ml.get_key_or_arr`, so a file
        // may declare a per-layer array instead of a scalar period.
        // ferrox carries ONE period for the whole model and
        // `metadata_u64_any` returns `None` for an array, so before this
        // gate the loader fell through to `default_swa_layout` and ran
        // the architecture's hardcoded layout in place of the file's --
        // silently, which is the failure this repo keeps finding.
        //
        // `capability::unsupported_feature_keys` used to refuse the key
        // outright, which stopped this AND stopped every legitimate
        // scalar. Now presence is fine and unreadability is not, so
        // both halves have to be tested: the scalar case is the plamo3
        // fixture in `tests/fixture_away_graphs.rs`, and this is the
        // array case.
        let pattern: [u32; 4] = [1, 0, 1, 0];
        let kvs: Vec<(&str, Kv)> = vec![
            ("general.architecture", Kv::Str("plamo3")),
            ("plamo3.block_count", Kv::U32(4)),
            ("plamo3.embedding_length", Kv::U32(64)),
            ("plamo3.attention.head_count", Kv::U32(1)),
            ("plamo3.attention.head_count_kv", Kv::U32(1)),
            ("plamo3.attention.key_length", Kv::U32(64)),
            ("plamo3.rope.freq_base", Kv::F32(10_000.0)),
            ("plamo3.attention.sliding_window", Kv::U32(3)),
            (
                "plamo3.attention.sliding_window_pattern",
                Kv::Arr32(&pattern),
            ),
        ];
        let file = open_metadata_gguf("swa_pattern_array", &kvs);
        match ModelConfig::from_gguf(&file) {
            Err(LoadError::UnsupportedFeature(arch, msg)) => {
                assert_eq!(arch, "plamo3");
                assert!(
                    msg.contains("not a scalar period"),
                    "the refusal must name what is wrong with the value: {msg}"
                );
            }
            other => panic!("an array-valued SWA pattern must refuse, got {other:?}"),
        }

        // And the scalar spelling of the same key still loads, or the
        // gate would be refusing the feature rather than the shape.
        let mut scalar = kvs;
        scalar.pop();
        scalar.push(("plamo3.attention.sliding_window_pattern", Kv::U32(2)));
        let file = open_metadata_gguf("swa_pattern_scalar", &scalar);
        let config = ModelConfig::from_gguf(&file).expect("a scalar period must load");
        assert_eq!(config.swa_pattern, Some(2));
        assert_eq!(config.layer_sliding_window(0), Some(3));
        assert_eq!(config.layer_sliding_window(1), None);
    }

    /// Baichuan is one `general.architecture` string covering two
    /// positional schemes, and llama.cpp picks between them on the layer
    /// count alone (`src/models/baichuan.cpp:11-14`, with its own "TODO:
    /// become GGUF KV parameter"). So the 13B is the MiniCPM case: no
    /// key to gate on and no tensor to miss.
    #[test]
    fn baichuan_13b_is_refused_because_it_uses_alibi_and_the_7b_is_not() {
        let thirteen_b = open_metadata_gguf(
            "baichuan13b",
            &[
                ("general.architecture", Kv::Str("baichuan")),
                ("baichuan.block_count", Kv::U32(40)),
            ],
        );
        match ModelConfig::from_gguf(&thirteen_b) {
            Err(LoadError::UnsupportedFeature(arch, msg)) => {
                assert_eq!(arch, "baichuan");
                assert!(msg.contains("ALiBi"), "{msg}");
                assert!(
                    msg.contains("40"),
                    "the refusal must name the layer count: {msg}"
                );
            }
            other => panic!("Baichuan-13B must be refused, got {other:?}"),
        }

        // The 7B rotates exactly as the generic decoder does, so it must
        // pass this gate. It still fails later, on the next missing
        // hparam, which is what proves the gate let it through.
        let seven_b = open_metadata_gguf(
            "baichuan7b",
            &[
                ("general.architecture", Kv::Str("baichuan")),
                ("baichuan.block_count", Kv::U32(32)),
            ],
        );
        match ModelConfig::from_gguf(&seven_b) {
            Err(LoadError::MissingHparam(key)) => assert_eq!(key, "baichuan.embedding_length"),
            other => panic!("Baichuan-7B must pass the ALiBi gate, got {other:?}"),
        }
    }

    /// The hyper-parameters a real Gemma-3 GGUF header carries for one
    /// size. `block_count` is the field llama.cpp's `LLM_TYPE_27B`
    /// switch reads (`gemma3.cpp:20-28`), so it is never a free
    /// parameter here.
    ///
    /// `linear_factor` adds the pair `conversion/base.py:1222-1230`
    /// writes from `rope_parameters["full_attention"]` -- and only from
    /// there: its own comment is "TODO: Handle sliding_attention
    /// similarly when models start implementing it", so the sliding
    /// layers get no scaling key at all.
    fn gemma3_config(
        tag: &str,
        n_layers: u32,
        hidden_dim: u32,
        n_heads: u32,
        head_dim: u32,
        linear_factor: Option<f32>,
    ) -> ModelConfig {
        let mut kvs: Vec<(&str, Kv)> = vec![
            ("general.architecture", Kv::Str("gemma3")),
            ("gemma3.block_count", Kv::U32(n_layers)),
            ("gemma3.embedding_length", Kv::U32(hidden_dim)),
            ("gemma3.attention.head_count", Kv::U32(n_heads)),
            ("gemma3.attention.head_count_kv", Kv::U32(n_heads)),
            ("gemma3.attention.key_length", Kv::U32(head_dim)),
            ("gemma3.attention.value_length", Kv::U32(head_dim)),
            // Global layers rotate at 1e6; the sliding ones fall back to
            // llama.cpp's `rope_freq_base_train_swa` default of 10000,
            // because `gemma3.cpp:11` reads only the BASE key.
            ("gemma3.rope.freq_base", Kv::F32(1_000_000.0)),
            ("gemma3.attention.sliding_window", Kv::U32(1024)),
            ("gemma3.attention.sliding_window_pattern", Kv::U32(6)),
        ];
        if let Some(factor) = linear_factor {
            kvs.push(("gemma3.rope.scaling.type", Kv::Str("linear")));
            kvs.push(("gemma3.rope.scaling.factor", Kv::F32(factor)));
        }
        ModelConfig::from_gguf(&open_metadata_gguf(tag, &kvs)).expect("gemma3 fixture must load")
    }

    /// The 27B attention scale reaches `ModelConfig`, and no other
    /// Gemma-3 size acquires one.
    ///
    /// `capability::attention_scale_override` is where the arithmetic is
    /// checked; this pins that the LOADER calls it with this file's own
    /// numbers. That step is the one that shipped broken: the function
    /// did not exist and `attention_scale` was the literal `None`, under
    /// a comment naming the exception. A helper nobody calls looks
    /// exactly like a fix.
    #[test]
    fn a_gemma3_27b_header_sets_the_attention_scale_and_no_smaller_size_does() {
        // Gemma-3-27B: 62 layers, n_embd 5376, 32 heads, head_dim 128.
        let big = gemma3_config("g3_27b_scale", 62, 5376, 32, 128, Some(8.0));
        let want = 1.0f32 / (5376.0f32 / 32.0).sqrt();
        let got = big
            .attention_scale
            .expect("Gemma-3-27B is llama.cpp's LLM_TYPE_27B");
        assert!(
            (got - want).abs() < 1e-7,
            "want 1/sqrt(168) = {want}, got {got}"
        );
        // The bug's magnitude: scores were sqrt(168/128) = 1.146x too
        // large without this.
        let kernel = 1.0f32 / 128.0f32.sqrt();
        assert!((kernel / got - (168.0f32 / 128.0).sqrt()).abs() < 1e-5);

        // Gemma-3-1B and -4B take llama.cpp's other branch, which is the
        // scale the attention kernels already apply. A `Some` here would
        // double-scale them.
        for (tag, n_layers, hidden, heads) in
            [("g3_1b_scale", 26, 1152, 4), ("g3_4b_scale", 34, 2560, 8)]
        {
            let cfg = gemma3_config(tag, n_layers, hidden, heads, 256, None);
            assert_eq!(
                cfg.attention_scale, None,
                "{tag} must keep the kernels' own 1/sqrt(head_dim)"
            );
        }
    }

    /// Gemma-3's declared linear scaling reaches the FULL-ATTENTION
    /// layers only, and the sliding ones rope unscaled.
    ///
    /// `gemma3.cpp:11` reads `LLM_KV_ROPE_FREQ_BASE_SWA` and nothing
    /// else, so `rope_freq_scale_train_swa` keeps its `1.0f` default
    /// (`src/llama-hparams.h:129`) while `get_rope_freq_scale`
    /// (`llama-model.cpp:2033-2035`) hands the trained scale to the full
    /// layers. The converter agrees: `conversion/base.py:1222-1230`
    /// takes the factor from `rope_parameters["full_attention"]` and
    /// writes nothing for the sliding half.
    ///
    /// ferrox folded the factor into ONE global `rope_freqs` vector, so
    /// Gemma-3-4B/12B/27B rotated five layers in six at `p/8`.
    #[test]
    fn gemma3_linear_scaling_reaches_the_full_layers_and_not_the_sliding_ones() {
        // Gemma-3-4B: 34 layers, head_dim 256, `rope_scaling: linear 8`.
        let cfg = gemma3_config("g3_4b_rope", 34, 2560, 8, 256, Some(8.0));
        let freqs = cfg
            .rope_freqs
            .as_ref()
            .expect("declared linear scaling must produce per-band divisors");
        assert!(
            freqs.full.iter().all(|f| (*f - 8.0).abs() < 1e-6),
            "full-attention layers divide every band by the trained factor: {:?}",
            freqs.full
        );
        let swa = freqs
            .swa
            .as_ref()
            .expect("gemma3 does not assign rope_freq_scale_train_swa, so 1.0 applies");
        assert!(
            swa.iter().all(|f| (*f - 1.0).abs() < 1e-6),
            "sliding layers rope at the raw position: {swa:?}"
        );
        assert_eq!(swa.len(), freqs.full.len(), "one divisor per rotated pair");

        // Period 6, last-dense (`capability::default_swa_layout`), so
        // layer 5 is the full-attention one and 0..=4 slide. The phase
        // matters: getting it wrong swaps which five-sixths are wrong.
        assert!(cfg.layer_sliding_window(0).is_some());
        assert!(cfg.layer_sliding_window(5).is_none());
        assert_eq!(cfg.layer_rope_freqs(0), Some(&[1.0f32; 128][..]));
        assert_eq!(cfg.layer_rope_freqs(5), Some(&[8.0f32; 128][..]));
        assert_eq!(cfg.layer_rope_theta(0), 10_000.0);
        assert_eq!(cfg.layer_rope_theta(5), 1_000_000.0);
        assert!(
            cfg.rope_freqs_vary_by_layer(),
            "the fused Metal stacks take one divisor slice for a whole run \
             and so must refuse this model"
        );

        // Gemma-3-1B declares no scaling at all -- the audited fixture,
        // and the reason this was invisible. Nothing to split, so no
        // per-layer set and no Metal refusal.
        let plain = gemma3_config("g3_1b_rope", 26, 1152, 4, 256, None);
        assert!(plain.rope_freqs.is_none());
        assert!(!plain.rope_freqs_vary_by_layer());
    }

    /// Gemma-2 is the counter-case, and it is why the SWA scale needs
    /// its own table rather than reusing `swa_rope_base_follows_model`.
    ///
    /// `gemma2.cpp:10-11` assigns BOTH `rope_freq_base_train_swa` and
    /// `rope_freq_scale_train_swa` from the model's trained values, so
    /// its sliding layers keep the declared scaling. Splitting them here
    /// would be the same bug pointed the other way.
    #[test]
    fn gemma2_sliding_layers_inherit_the_trained_rope_scale() {
        let cfg = ModelConfig::from_gguf(&open_metadata_gguf(
            "g2_rope",
            &[
                ("general.architecture", Kv::Str("gemma2")),
                ("gemma2.block_count", Kv::U32(26)),
                ("gemma2.embedding_length", Kv::U32(2304)),
                ("gemma2.attention.head_count", Kv::U32(8)),
                ("gemma2.attention.head_count_kv", Kv::U32(4)),
                ("gemma2.attention.key_length", Kv::U32(256)),
                ("gemma2.attention.value_length", Kv::U32(256)),
                ("gemma2.rope.freq_base", Kv::F32(10_000.0)),
                ("gemma2.attention.sliding_window", Kv::U32(4096)),
                ("gemma2.rope.scaling.type", Kv::Str("linear")),
                ("gemma2.rope.scaling.factor", Kv::F32(8.0)),
            ],
        ))
        .expect("gemma2 fixture must load");

        let freqs = cfg.rope_freqs.as_ref().expect("linear scaling declared");
        assert_eq!(
            freqs.swa, None,
            "gemma2.cpp:11 assigns rope_freq_scale_train_swa from the trained scale"
        );
        assert!(!cfg.rope_freqs_vary_by_layer());
        // Period 2, last-dense: layer 0 slides, layer 1 does not, and
        // both get the same divisors.
        assert!(cfg.layer_sliding_window(0).is_some());
        assert!(cfg.layer_sliding_window(1).is_none());
        assert_eq!(cfg.layer_rope_freqs(0), cfg.layer_rope_freqs(1));

        // The two tables really are different: this is the pair that
        // must not be collapsed into one.
        assert!(crate::capability::swa_rope_scale_follows_model("gemma2"));
        assert!(!crate::capability::swa_rope_scale_follows_model("gemma3"));
        for arch in ["olmo2", "laguna"] {
            assert!(
                crate::capability::swa_rope_base_follows_model(arch),
                "{arch} seeds the SWA base from the model"
            );
            assert!(
                !crate::capability::swa_rope_scale_follows_model(arch),
                "{arch} pins the SWA scale to 1.0 (olmo2.cpp:14, laguna.cpp:48)"
            );
        }
    }
}