Albor LLM Specification

Version: 0.6.0 Date: 2026-03-03 Status: Phase 3 — 350M Base Model Retraining (ALB-060 fix, v2 data) Author: Noah Gift / Pragmatic AI Labs

Albor (Spanish: “dawn”) — A sovereign Python code completion model trained from first principles using only the Sovereign AI stack. Python-only following the phi-1 playbook: maximum concentration on one language, distilled from Qwen3-Coder-Next (80B), then optimized through fine-tuning, merging, pruning, and quantization into a fast, local, zero-dependency code completion engine. The goal is twofold: produce a usable Python code assist model that runs anywhere Rust compiles, and identify + fix every gap in the stack that blocks end-to-end LLM development.

Latest milestone: 350M CUDA test training verified — 50 steps, loss 10.39→5.92 (best 5.53), checkpoint loads in realizar, all training stability contracts pass. First full training run failed (ALB-060: epochs=1 only ran 43/5000 steps). Fixed with C-TRAINCFG-001 contract + v2 config (67,977 sequences, 139M tokens, epochs=38). Qwen2.5-Coder-3B interim teacher validated for distillation. 24+ upstream gaps fixed across 8 sovereign stack components.

1. Objectives

1.1 Primary Goal

Train, distill, and optimize a 350M-parameter decoder-only transformer using exclusively the Sovereign AI stack:

• apr for training, distillation, merging, pruning, quantization, eval, export
• alimentar for data loading and preprocessing
• forjar for pipeline orchestration (DAG engine, multi-machine, state tracking)
• bashrs (Rash) for shell fragment validation in pipeline task resources
• repartir for distributed compute
• entrenar for the training engine (autograd, optimizers, checkpointing)
• trueno for SIMD/GPU tensor operations
• realizar for inference (teacher model, eval, serving)
• presentar for training visualization (TUI dashboards, experiment browser, WASM)
• batuta for orchestration, stack coordination, and falsification
• pv (provable-contracts) for design-by-contract verification of every kernel
• pmat for TDG scoring, compliance, fault pattern analysis, and coverage gaps
• certeza for three-tier test effectiveness (unit → property → formal)

1.2 Secondary Goal (Stack Validation)

Identify every implementation gap that blocks the primary goal. Fix each gap in the correct upstream component. The model is the proof; the stack improvements are the lasting value.

1.3 Multi-Stage Improvement Ladder

The model is not a single training run — it is iteratively improved through every post-training technique available in apr. Each stage exercises a different part of the stack, produces a benchmarked checkpoint, and may reveal new gaps.

Stage 1: Pre-train base model         → albor-base
Stage 2: Distill from Qwen3-Coder-Next → albor-distill
Stage 3: Instruction fine-tune (LoRA)  → albor-instruct
Stage 4: Merge with complementary model → albor-merged
Stage 5: Prune for efficiency          → albor-pruned
Stage 6: Quantize for deployment       → albor-q4

1.4 Target Use Cases

Primary: Sovereign Code Assist

A tiny, fast, zero-dependency code completion model that runs entirely locally. No API calls, no Python runtime, no telemetry, no cloud. Distillation from Qwen3-Coder-Next gives it coding capability far above what 350M parameters normally achieve.

Language: Python only. Following the phi-1 playbook — maximum concentration on a single language produces dramatically better results at small param counts than spreading tokens across many languages. A 350M model that completes Python well is more useful than a 350M model that completes 10 languages poorly.

What Albor is NOT: It is not a chat model, not an instruction follower, not a reasoning engine, not a polyglot code model. It is a fast, local Python code completion kernel — the kind of model that lives inside an editor extension and fires on every keystroke.

Secondary: Stack Demonstration & Teaching Artifact

The model exists equally to prove the Sovereign AI stack can train, distill, optimize, and serve an LLM end-to-end in pure Rust. The HuggingFace model card is a tour of the stack. The reproducibility protocol means anyone can retrain from scratch using only apr commands.

1.5 What Albor Builds

Albor is a project repo, not a library. It contains no production Rust code. All Rust changes happen upstream in the sovereign stack components. Albor drives the upstream work, validates it end-to-end, and produces the model.

1.5.1 What Lives in Albor (This Repo)

albor/
├── docs/
│   ├── specifications/albor-llm-spec.md    # This spec
│   ├── model-card.md                       # HuggingFace model card
│   └── falsification-report.md             # batuta falsify output
├── configs/
│   ├── train/
│   │   ├── pretrain-50m.yaml              # 50M: model arch + training (pipeline validation)
│   │   ├── pretrain-125m.yaml             # 125M: model arch + training (intermediate)
│   │   ├── pretrain-350m.yaml             # 350M: model arch + training (final)
│   │   ├── distill.yaml                   # Distillation config
│   │   └── finetune-lora.yaml             # LoRA fine-tuning config
│   ├── pipeline/
│   │   └── albor.yaml                      # THE manifest: infra + data + train + eval + publish
│   ├── dashboard/
│   │   └── albor-dashboard.yaml            # presentar dashboard (TUI + WASM)
│   └── data-mix.yaml                       # Data source weights + upsampling
├── contracts/
│   ├── knowledge-distillation-kernel-v1.yaml  # ALB-013
│   ├── bpe-tokenizer-kernel-v1.yaml           # ALB-014
│   ├── model-merging-kernel-v1.yaml           # ALB-015
│   ├── pruning-kernel-v1.yaml                 # ALB-016
│   └── gradient-accumulation-kernel-v1.yaml   # ALB-017
├── tests/
│   ├── falsify/                            # FALSIFY-ALBOR-001 through 009
│   ├── integration/                        # End-to-end pipeline tests
│   └── smoke/                              # Quick sanity checks (50M model)
├── state/                                  # (gitignored) forjar state + locks
│   ├── lambda/state.lock.yaml              # Per-machine resource state
│   ├── intel/state.lock.yaml
│   └── forjar.lock.yaml                    # Global pipeline state
├── data/                                   # (gitignored) Training data
├── checkpoints/                            # (gitignored) Model checkpoints
└── eval/                                   # (gitignored) Evaluation results

1.5.2 apr as Unified Entry Point

apr is the single CLI for all model operations. It delegates to sibling projects (entrenar, alimentar, realizar, etc.) under the hood. If a subcommand doesn’t exist yet, we file a GitHub issue, implement it in the correct upstream repo, wire it into apr, dogfood it in albor, and close the issue.

Design Principle: Plan/Apply Everywhere

Every apr subcommand that touches data, compute, or infrastructure follows a plan/apply contract inspired by Terraform and forjar:

plan   → Validate inputs, estimate cost, show what WILL happen. No side effects.
apply  → Execute the plan. Mutates state (files, models, infrastructure).

This is not optional. It is the unifying design principle of the CLI. Every expensive operation gets a free dry-run. Every destructive operation shows you what it will do before it does it. Users never commit GPU hours, disk space, or network bandwidth without seeing the plan first.

The contract:

1. apr <cmd> plan <config> — Parse config, validate paths, estimate resources (VRAM, disk, time, tokens), print a human-readable execution plan. Exit 0 if valid, exit 1 with diagnostics if not. No GPU, no writes, no network.
2. apr <cmd> apply <config> — Execute. Reads the same config, does the work. Can be interrupted and resumed.
3. apr <cmd> validate <config> — Alias for plan with --strict schema-only checking (no resource estimation). Fast enough for CI.

Why this matters for albor: Training a 350M model for 7 days on a 4090 is not something you retry casually. A config typo caught at plan time saves days. A VRAM overestimate caught at plan time prevents OOM crashes at step 15,000. Plan/apply turns “hope it works” into “prove it will work, then run it.”

Dispatch Table

apr <subcommand>
├── pipeline plan/apply      → forjar DAG engine (THE entry point — runs everything)
├── tokenize plan/apply      → aprender BPE tokenizer
├── train plan/apply         → entrenar TransformerTrainer
├── distill plan/apply       → entrenar + realizar (precompute + student training)
├── finetune plan/apply      → entrenar LoRA/QLoRA
├── eval plan/apply          → aprender eval harness
├── merge plan/apply         → entrenar SLERP/TIES/DARE
├── prune plan/apply         → entrenar WANDA/magnitude
├── quantize plan/apply      → entrenar Q4/Q8
├── export plan/apply        → entrenar SafeTensors/GGUF
├── publish plan/apply       → entrenar HuggingFace Hub
├── bench plan/apply         → realizar latency benchmarks
├── provision plan/apply     → forjar infrastructure convergence
├── experiment view/export   → presentar TUI + entrenar SQLite
└── monitor                  → presentar live TUI (reads training_state.json)

apr pipeline is the top-level command. It reads a single YAML manifest that describes infrastructure resources AND training tasks in one DAG. Forjar’s engine resolves dependencies (Kahn’s toposort), tracks state (BLAKE3 hashes), and dispatches each step — calling back into apr subcommands for ML tasks. Individual subcommands (apr train, apr eval, etc.) still work standalone for development and debugging.

Plan Output Format

Every plan subcommand prints a structured summary:

$ apr train plan configs/train/pretrain-350m.yaml

  Albor Train Plan
  ─────────────────────────────────────────────
  Model:        llama (24L, 1024H, 16A, 4KV)
  Parameters:   354,267,136 (~354M)
  Precision:    fp16 mixed
  ─────────────────────────────────────────────
  VRAM Budget:
    Weights       700 MB
    Optimizer   2,800 MB   (AdamW fp32 m+v)
    Gradients     700 MB
    Activations 9,200 MB   (grad ckpt, batch=8, seq=2048)
    Total      13,400 MB   (55.8% of 24,576 MB)
    Headroom   11,176 MB   ✓
  ─────────────────────────────────────────────
  Data:
    Train shards  data/tokenized/train/ (47 files, 8.2 GB)
    Val shards    data/tokenized/val/   (3 files, 410 MB)
    Tokenizer     models/albor-tokenizer/tokenizer.json ✓
    Vocab match   32,768 = model.vocab_size ✓
  ─────────────────────────────────────────────
  Training:
    Global batch  524,288 tokens (8 × 32 × 2048)
    Total tokens  10,000,000,000 (~10B)
    Total steps   19,073
    Warmup        2,000 steps (10.5%)
    Checkpoints   19 (every 1,000 steps)
    Disk est.     ~13.3 GB (19 × 700 MB)
  ─────────────────────────────────────────────
  Estimated wall time: 5.2 days on RTX 4090
  ─────────────────────────────────────────────
  ✓ Plan valid. Run `apr train apply configs/train/pretrain-350m.yaml` to start.

Forjar already does this (forjar plan -f albor.yaml). Entrenar has the TrainingPlan module (training_plan.rs) that mirrors forjar’s architecture. Albor’s job is to close the loop: every apr subcommand gets plan/apply, and every gap (ALB-XXX) that adds a new subcommand must implement both phases.

What Plan Validates Per Subcommand

1.5.3 Development Workflow: Issue-Driven Dogfooding

When albor hits a wall — a missing subcommand, a broken feature, a gap in a sibling project — the workflow is:

1. Hit wall       → apr <subcommand> doesn't exist or fails
2. File issue     → GitHub issue on correct repo (aprender, entrenar, alimentar, etc.)
3. Implement      → Fix upstream in the correct component
4. Wire into apr  → Add/update apr subcommand if needed
5. Dogfood        → Run the blocked albor pipeline step
6. Prove          → Tests pass, FALSIFY test passes, pmat comply check
7. Close issue    → Link to albor gap ID (ALB-XXX)

Every ALB-XXX gap in the gap register (§11) maps to a GitHub issue. The gap is not “closed” until the apr subcommand works end-to-end in the albor pipeline.

1.5.4 What Lives Upstream (Other Repos)

1.5.5 Where Quality Constraints Apply

The albor repo has no Rust code to cover. Its quality is measured by:

• Do the configs work? (integration tests)
• Do the FALSIFY tests pass? (end-to-end validation)
• Are the contracts complete? (pv status)
• Does the pipeline reproduce? (deterministic re-run)

1.6 Constraints

• Zero Python dependencies — Pure Rust from data to deployment
• Scientifically reproducible — Fixed seeds, versioned data, deterministic training
• Publicly auditable — All data, code, hyperparameters, and training logs published
• apr only — Every model operation uses an apr <subcommand>. Missing commands are gaps to implement.
• Plan/apply everywhere — Every apr subcommand implements plan (dry-run, no side effects) and apply (execute). No GPU time without a passing plan.
• One manifest, one DAG — apr pipeline plan/apply configs/pipeline/albor.yaml orchestrates the entire pipeline. No Makefiles, no shell scripts. Forjar’s DAG engine handles dependency resolution, state tracking, multi-machine dispatch, and resumability.
• bashrs linted — All shell fragments in forjar task resources are validated by bashrs (Rash). No unvalidated shell.
• No file over 500 lines — Applies to all code, scripts, configs, and contracts (not docs/specs)
• Provably correct — Every kernel has a YAML contract with falsification tests and Kani proofs
• pmat compliant — Upstream changes: TDG grade A, 95% coverage, 85% mutation score, zero SATD
• Falsifiable — Every claim in this spec has a concrete test that could disprove it

1.7 Sovereign Stack vs. Standard ML Stack

Most LLM training stacks depend on a deep tower of NVIDIA and Python libraries:

Standard ML Stack              Sovereign Stack (albor)
─────────────────              ──────────────────────
Python                         Rust (no Python runtime)
PyTorch / JAX                  entrenar (training engine)
cuDNN                          trueno PTX kernels + cuBLAS FFI
NCCL                           (not needed — single GPU)
torch.distributed              repartir (stretch goal)
Weights & Biases               presentar + renacer tracing
HuggingFace Transformers       realizar (inference)

What each replaced component does — and why we don’t use it:

The GPU interface: The sovereign stack talks to NVIDIA hardware through two interfaces only:

1. CUDA Driver API (libcuda.so) — Memory allocation, kernel launch, stream management, device queries. This is the lowest stable NVIDIA API. trueno binds it directly via Rust FFI — no CUDA Runtime API (libcudart) dependency.

2. cuBLAS (libcublas.so) — Matrix multiplication (GEMM). The only NVIDIA library used for compute. trueno wraps it with a safe Rust API (CublasHandle, CublasGemm) that enforces correct argument order at the type level. cuBLAS replaced hand-written PTX GEMMs in ALB-075, improving throughput from 890 tok/s to 6,700 tok/s (7.5x).

What this means in practice: The entire training binary is a single statically-linked Rust executable (~15 MB). It has no Python interpreter, no pip packages, no conda environment, no Docker container, no version conflicts between PyTorch and CUDA toolkit. cargo build --release produces a binary that runs training. The only runtime dependencies are libcuda.so (NVIDIA driver) and libcublas.so (ships with the driver).

2. Hardware Inventory

2.1 Machine: lambda (Threadripper)

2.2 Machine: intel (Mac Pro 2019 chassis, Linux)

2.3 Network

• SSH connectivity (ssh intel) with ControlMaster multiplexing (forjar FJ-252)
• LAN bandwidth assumed ≥1 Gbps

2.4 Key Insight: 300 GB RAM Enables CPU-Based Teacher Inference

The intel box’s 300 GB RAM fundamentally changes the distillation architecture. Qwen3-Coder-Next (80B params) fits entirely in CPU RAM:

No quantization-induced quality loss needed. The teacher runs at full fp16 precision, producing the highest-quality soft targets for distillation.

3. Model Architecture

3.1 Architecture: LLaMA-Style Decoder-Only Transformer

entrenar’s transformer is a pre-norm LLaMA-style architecture with RMSNorm, SwiGLU FFN, Grouped-Query Attention, and RoPE. This is hardcoded in the Transformer struct — we configure it via YAML, we don’t build it from scratch.

3.2 Progressive Model Sizing

To validate the pipeline quickly, we train progressively larger models. Each gets its own YAML config file (see §6.2 for full config format).

The 50M model proves the entire stack works end-to-end before committing days of GPU time to the 350M run.

3.3 VRAM Budget (fp16 mixed precision, RTX 4090)

Speculative estimates (pre-dogfooding):

Actual measurements (from ALB-040 dogfooding with CudaTransformerTrainer):

The GPU-resident CudaTransformerTrainer keeps all 24 blocks in VRAM (weights + AdamW states ≈ 5 GB) plus a shared workspace for activations (~10-12 GB). This is tighter than the speculative estimate because the shared workspace includes attention score matrices that scale as O(heads × seq² × batch). Batch size is fixed at 4. Note: gradient_accumulation is set to 1 for the v2 config, though per-block CPU gradient accumulation is now fully implemented via PerBlockGradientAccumulator (D2H download, CPU averaging, H2D upload). See §6.4 for detailed breakdown.

4. Distillation Teacher: Qwen3.5-35B-A3B

4.1 Teacher Model Profile

4.2 Why This Teacher

• Apache 2.0: Legally clean for distillation, no license contamination
• 35B knowledge at 3B cost: MoE activates only 8+1 experts per token. Inference FLOP budget matches a dense 1.8B model, but the 256 experts collectively encode 35B parameters of knowledge. Soft targets are far richer than a dense 3B teacher.
• Fits on a single 4090: At Q4 quantization, weights occupy ~17.5 GB. With activations and KV cache (only 10 full-attention layers need KV cache), total VRAM is ~18.3 GB — leaving 5.7 GB headroom on 24 GB.
• Coding focus: Distilled student inherits strong code capabilities, making it competitive on HumanEval/MBPP — benchmarks where tiny models normally fail.
• realizar already supports most of the architecture: Gated DeltaNet linear attention (GH-278), SwiGLU FFN, GQA, hybrid layer_types config, and MoE routing (CapacityFactorRouter, PowerOfTwoChoicesRouter) all exist. The missing pieces are expert weight loading and dispatch integration.
• Novel architecture (DeltaNet + MoE): Exercising realizar’s model loading on a non-standard architecture is exactly the kind of gap-finding that validates the stack.

4.2.1 VRAM Budget (Q4, batch=1, seq=2048)

4.2.2 Realizar MoE Readiness Assessment

Estimated new code: ~300-400 lines in realizar for full MoE inference.

4.3 Distillation Architecture

Primary path: GPU-resident teacher inference on lambda (RTX 4090). The 35B model at Q4 fits in 18.3 GB VRAM — teacher inference and logit caching run on the same machine as student training.

┌─────────────────────────────────────────────────────────────────────────┐
│  lambda (RTX 4090, 24 GB)                                              │
│                                                                         │
│  Phase 1: Pre-compute teacher logits (GPU, ~18.3 GB)                   │
│  ┌──────────────────────────┐     Parquet shards      ┌──────────────┐ │
│  │ Qwen3.5-35B-A3B (Q4)    │ ──────────────────────► │ teacher_logits│ │
│  │ realizar MoE inference   │    top-k=128 logits     │ ~50-100 GB   │ │
│  │ 18.3 GB VRAM             │                          └──────────────┘ │
│  └──────────────────────────┘                                           │
│                                                                         │
│  Phase 2: Train student (GPU, ~5 GB)                                   │
│  ┌──────────────────────────┐     ┌─────────────────────────────────┐  │
│  │ Student: albor-350M      │ ◄── │ Pre-computed logits + train data │  │
│  │ KD loss + CE loss        │     │ (loaded from disk at GPU speed)  │  │
│  │ entrenar distill         │     └─────────────────────────────────┘  │
│  └──────────────────────────┘                                           │
└─────────────────────────────────────────────────────────────────────────┘

Fallback path: If GPU VRAM is tight (teacher + student simultaneously), pre-compute logits on CPU. Intel box (300 GB RAM) can run the 35B model at Q4 (~18 GB RAM) or Q8 (~35 GB) with ~5-15 tok/s throughput.

4.4 Pre-Computed Logits Strategy

Teacher and student do NOT run simultaneously. We pre-compute teacher logits offline, then train the student from cached logits at full GPU speed:

1. Lambda runs Qwen3.5-35B-A3B inference (Q4, GPU) on all training data
2. Teacher top-k logits (k=128) saved as sharded Parquet via alimentar
3. Student training loads pre-computed logits from disk — no teacher in VRAM
4. Sequential phases = no VRAM contention

# Step 0: Plan — check teacher fits, estimate logit disk usage
apr distill plan configs/train/distill.yaml

# Step 1: Pre-compute teacher logits on lambda GPU (Q4, ~18.3 GB)
apr distill apply configs/train/distill.yaml --stage precompute

# Step 2: Train student on lambda using pre-computed logits (~5 GB)
apr distill apply configs/train/distill.yaml --stage train --seed 42

Estimated teacher throughput (Qwen3.5-35B-A3B):

4.5 Distillation Data Budget

Recommended: Start with the local ground truth corpora (~50-100M raw tokens) plus curated hard examples from StarCoder Python (~400M tokens) for ~500M total. The ground truth corpora should be distilled first — they are our highest quality data and benefit most from teacher knowledge. Scale to 2B with broader StarCoder data if benchmarks justify the compute. Python-only focus means all teacher compute goes toward the language we care about.

4.6 Fallback Teacher: Qwen2.5-Coder-3B

If ALB-010 (MoE inference in realizar) proves harder than estimated, we fall back to Qwen2.5-Coder-3B as a dense teacher:

Why this is the fallback, not the primary:

• Dense 3B has ~10x less knowledge capacity than 35B MoE
• Weaker code capabilities → lower distillation quality ceiling
• Soft targets less informative for the student

Why it’s still viable:

• Already supported by realizar’s Qwen2 architecture loader (no MoE/DeltaNet)
• apr distill --stage precompute verified working with 3B teacher (2026-03-03)
• CPU precompute feasible on lambda box (~12 GB RAM)
• 8.6x compression ratio is in the sweet spot for KD

Config: configs/train/distill-qwen3b.yaml — teacher: Qwen2.5-Coder-3B, student: albor-base-350m, temperature=4.0, alpha=0.5, LoRA rank 16.

4.7 ALB-010 Implementation Status: MoE Inference in Realizar

Status: MERGED — Steps 1-5b merged to main (PR #133, squash-merged).

Step 1: Expert weight types + loading — DONE

• MoeExpertWeights struct in gpu/scheduler/types.rs (45 files updated)
• Fields: gate_weight, expert_gate_up, expert_down, shared_{gate,up,down}
• GpuModelConfig extended with num_experts, num_experts_per_tok, expert_intermediate_size

Step 2: Router forward — DONE (moe_dispatch.rs)

• moe_route(): softmax (max-subtracted) → top-k selection → renormalize
• 3 contract-derived tests pass: stability, uniform routing, order preservation

Step 3: Expert dispatch — DONE (moe_dispatch.rs)

• expert_swiglu(): per-expert down(SiLU(gate(x)) * up(x))
• moe_forward_token(): routes to k experts + shared expert, weighted sum
• 2 contract-derived tests pass: shared expert always active, uniform routing averages

Step 4: Integration into forward pass — DONE

• All 5 forward block variants integrated: forward_block_refcell, forward_block_single, forward_block_incremental, forward_block_incremental_optimized, forward_block_idx
• MoE path activates when block.moe_experts.is_some()
• Multi-token forward_block_idx loops per token (MoE routes independently per token)
• 15,053 total tests pass (0 failures)

Remaining: Safetensors weight loading

• Map HuggingFace tensor names (model.layers.{N}.mlp.experts.*) to MoeExpertWeights
• Fuse individual expert gate/up projections into expert_gate_up tensor
• Blocked on: model download (Qwen3.5-35B-A3B, ~70 GB)

4.8 Provable Contracts for MoE Inference

Two design-by-contract YAMLs written and validated (pv validate PASS) before implementation begins, per engineering discipline Rule #6:

contracts/moe-router-v1.yaml (Router forward):

• 4 equations: router_logits, softmax_normalization, topk_selection, weight_renormalization
• 6 invariants: softmax_valid, topk_ordered, renorm_sum_one, expert_count, index_bounds, deterministic
• 5 falsification tests: softmax stability with large logits, top-8 correctness, renorm ordering, zero gate weight, shape mismatch rejection
• 1 Kani harness (stub_float strategy for symbolic f32)

contracts/moe-expert-dispatch-v1.yaml (Expert dispatch):

• 5 equations: expert_swiglu, routed_output, shared_expert, moe_output, fused_gate_up_unfuse
• 6 invariants: expert_output_shape, weighted_sum_preserves_shape, shared_expert_always_active, expert_independence, unfuse_covers_all, numerical_stability
• 7 falsification tests: single-expert routing, uniform routing, unfuse round-trip, shared expert unconditional, bounds check, finite outputs, dense FFN equivalence
• 2 Kani harnesses (bounded_int strategy)

Performance characteristics (from docs/specifications/training-performance.md §6.19):

• 28 GEMMs per token per MoE layer (vs 3 for dense FFN)
• Expert GEMMs are tiny ([2048, 512]) — memory-bandwidth bound at batch=1
• Router overhead negligible vs expert computation
• Estimated teacher throughput: 50-100 tok/s on RTX 4090 at Q4

4.9 Qwen3.5-35B-A3B Tensor Name Mapping

Architecture class: Qwen3_5MoeForConditionalGeneration (model_type: qwen3_5_moe). All layer tensors use model.language_model.layers.{L} prefix (multimodal wrapper).

MoE Expert Tensors (packed per-layer, not per-expert):

Key architectural detail: The shared expert output is scaled by sigmoid(shared_expert_gate(x)) before adding to the routed expert sum. This was discovered from the HuggingFace source (Qwen3_5MoeSparseMoeBlock) and added to MoeExpertWeights.shared_expert_gate_weight in realizar.

Expert weights are packed: Unlike per-expert indexing (experts.{E}.gate_proj), the main model stores all 256 experts in bulk tensors (experts.gate_up_proj). The MTP (multi-token prediction) head uses per-expert indexing. Realizar handles the packed format directly in MoeExpertWeights.expert_gate_up.

5. Training Data

5.1 Data Philosophy

• All datasets either locally owned (MIT/Apache 2.0) or publicly available with permissive licenses
• Local-first: Sovereign ground truth corpora are our highest-quality data — curated, tested, type-annotated, and owned. They are upsampled to punch above their token weight.
• Exact download URLs, versions, and SHA-256 hashes recorded for all external data
• Preprocessing pipeline is deterministic (fixed seed, recorded transforms)
• Quality validated by alimentar quality check

5.2 Data Mix (Target: ~10B tokens)

Current status (2026-03-05): v3 dataset in preparation — 2M Python files from codeparrot-clean (~4.4B tokens raw, ~5.3B pretokenized at seq_len=1024). v2 dataset had only 139M tokens (67,977 sequences × 2048), which is 0.9% of Chinchilla-minimum for 350M params. v3 provides sufficient data for 1B+ token training runs. See §5.4.2 for the v3 pipeline.

Following the phi-1 playbook: maximum concentration on Python. phi-1 proved that a small model (1.3B) with focused data and distillation can hit 50% HumanEval — outperforming models 10x its size trained on diluted multi-language corpora.

Key insight from phi-1: Data quality matters more than quantity at small param counts. A 350M model trained on 1B tokens of textbook-quality code can outperform a 350M model trained on 100B tokens of raw GitHub scrapes. We have ~71K curated Python files locally — this is our unfair advantage.

Total: ~10B tokens. Chinchilla-optimal for 350M params is ~7B; we slightly overtrain for benchmark performance (common practice in SmolLM, Phi-1.5).

Python concentration: 80% of training data is Python or Python-adjacent (code, textbooks, docs). The remaining 20% (FineWeb-Edu) provides general language understanding for docstrings, comments, and natural language prompts.

5.2.1 Local Ground Truth Corpora (Tier 1 — Upsampled)

These are our “textbook-quality” data — the phi-1 equivalent. Every file has been curated, tested to 98%+ coverage, and validated by CI. They are upsampled 10x during training because their per-token teaching signal is 10-100x higher than raw GitHub code.

Why upsampling works: phi-1’s “textbook” data was <10% of total tokens but had outsized impact on HumanEval. Our ground truth corpora share the same properties: clean types, complete docstrings, tested correctness, educational structure. The model sees these examples multiple times, reinforcing correct patterns over noisy GitHub code.

depyler corpus is uniquely valuable: Every Python function in the depyler corpus was validated by a transpiler — it has clear types, clean control flow, and provably correct semantics. The tdd-book covers 27 stdlib modules (json, datetime, collections, itertools, os, pathlib, re, etc.) with property-based tests. This teaches the model Python’s standard library idioms at a depth no scraped dataset matches.

5.2.2 Local ML Framework Code (Tier 2)

Large, high-quality Python codebases from our local repos. Not upsampled — used at natural frequency for pattern diversity.

5.2.3 Pre-Built Local Datasets

5.2.4 Data Sourcing Summary

Local owned data (~71K files, ~1-2M lines):
├── Tier 1: Ground truth corpora (17,588 files) → upsampled 10x
├── Tier 2: ML framework code   (53,568 files) → natural frequency
└── Pre-built: Parquet + JSON manifests

External data (HuggingFace, ~8B tokens):
├── StarCoder Python subset     (~4B tokens)   → bulk diversity
├── FineWeb-Edu                 (~2B tokens)   → general language
├── Python textbooks/tutorials  (~1B tokens)   → educational code
└── Python docs + PEPs + SO     (~1B tokens)   → API knowledge

Sovereign data advantage: 20% of training tokens come from data we own, curate, and can improve. Unlike scraped web data, we know the provenance, license, and quality of every file. If benchmarks reveal weaknesses in specific Python patterns, we can add targeted examples to our ground truth corpora and retrain — a feedback loop no public-dataset-only approach can match.

5.3 Fill-in-the-Middle (FIM) Training

Code completion requires fill-in-the-middle capability, not just left-to-right generation. During training, a fraction of code sequences are transformed using the PSM (Prefix-Suffix-Middle) format:

<fim_prefix>def fibonacci(n):<fim_suffix>    return fib_sequence<fim_middle>
    fib_sequence = [0, 1]
    for i in range(2, n):
        fib_sequence.append(fib_sequence[-1] + fib_sequence[-2])

Gap ALB-018: FIXED — alimentar fim supports PSM/SPM transforms. Verified: alimentar fim mixed.parquet -o out.parquet --rate 0.5 --format psm --seed 42 produces correct FIM-encoded sequences. Used in v2 data pipeline.

This is critical — without FIM, the model is a text generator, not a code completion engine.

5.4 Data Pipeline

# ── Step 1: Ingest local ground truth corpora (Tier 1 — highest quality) ──
alimentar import local ../depyler/examples/ ../depyler/tdd-book/tests/ \
  --lang python --output ./data/local/depyler.parquet
alimentar import local ../hf-ground-truth-corpus/ \
  --lang python --output ./data/local/hf-gtc.parquet
alimentar import local ../jax-ground-truth-corpus/ \
  --lang python --output ./data/local/jax-gtc.parquet
alimentar import local ../vllm-ground-truth-corpus/ \
  --lang python --output ./data/local/vllm-gtc.parquet

# ── Step 2: Ingest local ML framework code (Tier 2) ──
alimentar import local \
  ../huggingface-fine-tuning/ ../llms-with-huggingface/ ../HF-Hub-Ecosystem/ \
  ../pytorch/ ../vllm/ ../databricks-data-engineering/ \
  ../algorithm-competition-corpus/ ../coursera-stats/ ../cuda-python/ \
  --lang python --output ./data/local/ml-frameworks.parquet

# ── Step 3: Download external data (on intel — 300GB RAM) ──
alimentar import hf bigcode/starcoderdata --lang python --output ./data/starcoder-python/
alimentar import hf HuggingFaceFW/fineweb-edu --output ./data/fineweb-edu/

# ── Step 4: Quality validation ──
alimentar quality check ./data/local/ --profile ml-training
alimentar quality check ./data/starcoder-python/ --profile ml-training
alimentar quality check ./data/fineweb-edu/ --profile ml-training

# ── Step 5: Filter, dedup, shard ──
alimentar filter ./data/starcoder-python/ --lang python --min-tokens 32 --max-tokens 8192 \
  --dedup --output ./data/processed/starcoder-python.parquet
alimentar convert ./data/fineweb-edu/ ./data/processed/fineweb-edu.parquet

# ── Step 6: Build training mix with upsampling ──
alimentar mix \
  --input ./data/processed/starcoder-python.parquet --weight 0.40 \
  --input ./data/local/depyler.parquet --weight 0.025 --upsample 10 \
  --input ./data/local/hf-gtc.parquet --weight 0.025 --upsample 10 \
  --input ./data/local/jax-gtc.parquet --weight 0.025 --upsample 10 \
  --input ./data/local/vllm-gtc.parquet --weight 0.025 --upsample 10 \
  --input ./data/local/ml-frameworks.parquet --weight 0.10 \
  --input ./data/processed/fineweb-edu.parquet --weight 0.20 \
  --input ./data/processed/textbooks.parquet --weight 0.10 \
  --input ./data/processed/python-docs.parquet --weight 0.10 \
  --output ./data/mixed/ \
  --seed 42 --shuffle

# ── Step 7: Record provenance ──
alimentar provenance ./data/mixed/ --output ./data/provenance.json

Gap ALB-019: FIXED — alimentar import local expects data files (CSV/JSON/Parquet), not source code directories. Workaround: scripts/source-to-parquet.py converts Python source repos to Parquet with the Tier 1 schema (file, source, text columns). Used for all Tier 2 imports.

Gap ALB-020: FIXED — alimentar mix supports weighted proportional sampling. Syntax: alimentar mix file1.parquet:10.0 file2.parquet:1.0 -o out.parquet.

5.4.1 Actual Pipeline (v2 Dataset — 2026-03-03)

The pipeline below produced the v2 dataset (139M tokens, 67,977 sequences):

# ── Step 1: Convert Tier 2 repos to Parquet (alimentar can't read source dirs) ──
for repo in pytorch hf-repos mlflow vllm-full tgi algo-corpus cuda-python llms-with-hf; do
    python3 scripts/source-to-parquet.py ~/src/$repo $repo data/parquet/tier2/$repo.parquet
done
# Result: 28,553 Python files across 8 repos

# ── Step 2: Mix Tier 1 (10x) + Tier 2 (1x) ──
alimentar mix \
  data/parquet/depyler/shard_0000.parquet:10.0 \
  data/parquet/hf-ground-truth/shard_0000.parquet:10.0 \
  data/parquet/jax/shard_0000.parquet:10.0 \
  data/parquet/vllm/shard_0000.parquet:10.0 \
  data/parquet/tier2/pytorch.parquet:1.0 \
  data/parquet/tier2/hf-repos.parquet:1.0 \
  data/parquet/tier2/mlflow.parquet:1.0 \
  data/parquet/tier2/vllm-full.parquet:1.0 \
  data/parquet/tier2/tgi.parquet:1.0 \
  data/parquet/tier2/algo-corpus.parquet:1.0 \
  data/parquet/tier2/cuda-python.parquet:1.0 \
  data/parquet/tier2/llms-with-hf.parquet:1.0 \
  -o data/staging/mixed-expanded.parquet --seed 42
# Result: 45,420 mixed rows

# ── Step 3: Apply FIM (50% PSM) ──
alimentar fim data/staging/mixed-expanded.parquet \
  -o data/staging/mixed-expanded-fim.parquet --rate 0.5 --format psm --seed 42
# Result: 45,420 rows with ~50% FIM-encoded

# ── Step 4: Pretokenize into 2048-length sequences ──
python3 scripts/pretokenize.py \
  --input data/staging/mixed-expanded-fim.parquet \
  --tokenizer models/albor-tokenizer-v2/tokenizer.json \
  --seq-len 2048 \
  --output data/pretokenized-2048-v2/train/train.parquet
# Result: 67,977 sequences × 2048 = 139,218,944 tokens (191 MiB)

# Validation set: reuse v1
cp data/pretokenized-2048/val/val.parquet data/pretokenized-2048-v2/val/val.parquet

5.4.2 v3 Dataset Pipeline — codeparrot-clean (2026-03-05)

The v3 dataset scales from 139M to ~5.3B tokens using codeparrot/codeparrot-clean (5M Python files on HuggingFace, no gating). Quality filtered and pretokenized at seq_len=1024 for the 350M model’s max_position_embeddings.

# Step 1: Stream and filter from HuggingFace (2M files, ~8 min)
python3 scripts/download-codeparrot.py \
  --output /mnt/nvme-raid0/albor-data/codeparrot-clean/ \
  --max-rows 2000000
# Filters: skip autogenerated, alpha_frac < 0.25, files > 100KB, < 50 chars
# Result: 2,000,000 files in 20 shards (6.1 GB), ~4.4B raw tokens est.

# Step 2: Pretokenize at seq_len=1024 (streaming shard-by-shard)
python3 scripts/pretokenize.py \
  --input /mnt/nvme-raid0/albor-data/codeparrot-clean/ \
  --tokenizer models/albor-tokenizer-v2/tokenizer.json \
  --seq-len 1024 \
  --output data/pretokenized-1024-v3/train/ \
  --text-column text --shard-output
# Result: ~5.2M sequences × 1024 = ~5.3B tokens in 20 output shards

# Validation set: reuse v1 (814 sequences)

5.5 Tokenizer

Existing capability: aprender::text::tokenize::BpeTokenizer with full train() / encode() / decode() support. entrenar::tokenizer::BPETokenizer provides the training-pipeline integration.

# Plan: validate inputs, estimate vocab training time
apr tokenize plan \
  --input ./data/processed/*.parquet \
  --vocab-size 32768 \
  --algorithm bpe \
  --output ./models/albor-tokenizer/

# Apply: train the tokenizer
apr tokenize apply \
  --input ./data/processed/*.parquet \
  --vocab-size 32768 \
  --algorithm bpe \
  --output ./models/albor-tokenizer/ \
  --seed 42

Gap ALB-001: Verify apr tokenize plan/apply exists as a CLI subcommand. If not, wire aprender::text::tokenize::BpeTokenizer::train() into apr with the plan/apply contract (see §1.5.2).

6. Training Configuration

6.1 Optimizer & Schedule

6.2 Training Config: configs/train/pretrain-350m-v2.yaml

A single YAML file defines everything — model architecture and training hyperparameters. This is the industry standard (Axolotl, torchtune, HuggingFace Trainer). One file, one truth. apr train validate lints it before GPU time.

Current config (v2 — expanded dataset, ALB-066 gradient_accumulation=1):

# configs/train/pretrain-350m-v2.yaml — Albor 350M with expanded dataset
# C-TRAINCFG-001: steps_per_epoch=16994 >= max_steps=5000

model:
  path: "."                                  # From scratch (random init)
  mode: transformer
  architecture:
    hidden_size: 1024                       # d_model
    num_hidden_layers: 24
    num_attention_heads: 16                 # d_head = 64
    num_key_value_heads: 4                  # GQA 4:1 ratio
    intermediate_size: 4096                 # SwiGLU FFN (gate + up + down)
    vocab_size: 32768                       # ByteLevel BPE (v2 tokenizer)
    max_position_embeddings: 1024           # Context length (2048 OOM'd on 4090)
    rms_norm_eps: 1.0e-5

data:
  train: "data/pretokenized-2048-v2/train/" # Expanded v2 dataset (68K sequences)
  val: "data/pretokenized-2048/val/"
  batch_size: 4                             # Micro-batch (batch=8 OOM'd)
  seq_len: 1024
  tokenizer: "models/albor-tokenizer-v2/tokenizer.json"
  input_column: "input_ids"                 # Pre-tokenized: List<u32> column

optimizer:
  name: "adamw"
  lr: 3.0e-4
  beta1: 0.9
  beta2: 0.95
  weight_decay: 0.1

training:
  mode: "causal_lm"
  epochs: 1                                 # C-TRAINCFG-001: steps_per_epoch=16994 >= 5000
  # grad_clip: 1.0                           # ALB-067: disabled (CPU-side L2 norm bottleneck)
  lr_scheduler: "cosine"
  warmup_steps: 500                         # 10% of max_steps (C-TRAINCFG-001)
  gradient_accumulation: 1                  # ALB-066: per-sequence optimizer (no true accum in CUDA)
  mixed_precision: "fp16"
  output_dir: "./checkpoints/albor-base-350m-v2"
  save_interval: 25
  max_steps: 5000

Legacy v1 config (pretrain-350m.yaml) used 22K sequences with gradient_accumulation: 128 and epochs: 117 — see ALB-060 for why epochs: 1 was fatal with the original data size.

Note on YAML numeric formatting: YAML supports underscore notation natively (32_768, 1_000_000) for human-readable large numbers. All albor configs use this convention. For shorthand like 10B or 512K, see gap ALB-021.

6.3 Training Workflow (Plan/Apply)

# Step 1: Plan — validate config, estimate VRAM, show execution plan (no GPU)
apr train plan configs/train/pretrain-350m.yaml

# Step 2: Apply — execute the training run
apr train apply configs/train/pretrain-350m.yaml --seed 42

# Step 3: Resume if interrupted (apply with --resume)
apr train apply configs/train/pretrain-350m.yaml \
  --resume checkpoints/albor-base-350m/checkpoint-step-5000.json \
  --seed 42

Plan phase (apr train plan):

• Schema validation: required keys, correct types, valid enum values
• Architecture sanity: hidden_size divisible by num_attention_heads, num_kv_heads divides num_attention_heads
• VRAM budget: computes model size + optimizer + activations, warns if > GPU capacity
• Data paths: confirms train: and val: directories exist with Parquet/tokenized shards
• Tokenizer: loads tokenizer, checks vocab size matches model.vocab_size
• Time estimate: estimated wall time based on model size and hardware
• Prints structured plan summary (see §1.5.2 for output format)
• No GPU, no writes, no network. Runs on CPU in seconds.

Apply phase (apr train apply):

• Reads the same YAML, builds a random-initialized Transformer with the model: section architecture, runs the causal LM training loop via entrenar
• Checkpoints every save_interval steps — resumable on crash
• No Rust code needed — just one config file

apr train validate is an alias for apr train plan --strict — schema-only checking without resource estimation. Fast enough for CI.

6.4 GPU-Resident Training (CudaTransformerTrainer)

The CudaTransformerTrainer (ALB-040) keeps all 24 transformer blocks GPU-resident, reducing PCIe transfers from ~16K/step to exactly 3:

Transfer 1 (H2D): embedding hidden states   ~S×H×4 bytes
Transfer 2 (D2H): logits for cross-entropy  ~S×V×4 bytes
Transfer 3 (H2D): grad_logits to GPU        ~S×V×4 bytes

Each CudaTransformerBlock holds its own weights, AdamW optimizer states (m + v), and shares a CudaGradWorkspace for forward/backward activation buffers. The per-block interleaved backward+optimizer pattern overwrites the shared workspace each layer — memory cost is O(1 block), not O(24 blocks) for activations.

VRAM budget (actual, RTX 4090 24GB):

At seq_len=512, batch=4: fits comfortably (~18 GB used). At seq_len=1024, batch=4: fits (~19.5 GB used). At seq_len=2048, batch=4: OOM at LM head alloc (logits [4,2048,32768] too large). At seq_len=2048, batch=8: OOM at block 21 upload.

Dogfooding results:

ALB-060: Training Configuration Epoch/Step Mismatch (Critical)

The first 350M full training run (2026-03-02) ran only 43 of 5000 steps because epochs: 1 caps total steps to floor(num_sequences / batch_size / grad_accum). With 22,079 sequences, batch=4, accum=128: steps_per_epoch = 43. Warmup (2000 steps) never completed — LR peaked at 6.45e-6 vs target 3e-4. Loss stayed flat at ~10.39 for all 43 steps (never exited warmup). Root cause: no pre-flight algebraic validation of epoch/step consistency.

Fix: C-TRAINCFG-001 contract (contracts/training-config-kernel-v1.yaml) + epochs: 117 for v1 data, or v2 config (pretrain-350m-v2.yaml) with expanded dataset (67,977 sequences, epochs: 38, warmup_steps: 500).

Training stability contracts verified (ALB-044, ALB-059, ALB-060):

• C-EMBED-GRAD-001: Activation gradient clipped at GPU→CPU boundary
• C-HYPERPARAMS-001: All optimizer params flow from YAML config
• C-BUFSIZE-001: Buffer sizes algebraically verified (ALB-043 fix)
• C-GRADFLOW-001: All trainable parameters receive gradients (ALB-038 fix)
• C-GEMMARGS-001: GEMM backward constructor args match documented order (ALB-059 fix)
• C-GPUINIT-001: Optimizer states zero-initialized, not cuMemAlloc garbage (ALB-059 fix)
• C-STREAMSYNC-001: stream.synchronize() before any D2H transfer reading kernel output (ALB-065 fix)
• C-LOSSSCALE-001: fp16 loss scaling excluded from f32 backward path (ALB-072 fix)
• C-SELP-001: PTX selp_f32 argument order verified in all kernels (ALB-069, ALB-073 fixes)
• C-EVALBUF-001: eval_single_sequence truncates to max_seq_len before GPU forward (ALB-074 fix)
• C-GPUINIT-001: All optimizer m/v buffers zero-initialized (ALB-059 fix)
• C-LOSSSCALE-001: fp16 loss scaling excluded from GPU backward (all backward uses f32; scaling causes overflow) (ALB-072 fix)
• C-CUBLAS-NOTENCORE-001: cuBLAS uses CUBLAS_DEFAULT_MATH (no tensor cores) — tensor core algorithms produce NaN for transposed backward GEMMs at ~1e5 gradient magnitude (ALB-077 fix)

6.5 Checkpointing Strategy

Checkpoint interval rationale (v3): save_interval: 1000 balances crash recovery (~8.7min max lost work at 525ms/step) against I/O overhead (~3s per checkpoint write vs ~525s between checkpoints = 0.6% overhead). With automatic pruning, disk usage stays constant regardless of training length. For the 250K-step v3 run (~1.5 days at 7,579 tok/s), this yields 250 checkpoint events with ~8.4 GB steady-state disk.

6.6 Experiment Tracking & Training Monitoring

entrenar has a full monitoring stack built in, and presentar provides rich terminal visualization. Albor uses both — no external tools (no W&B, no MLflow, no TensorBoard). Sovereign monitoring, sovereign visualization.

6.6.1 Monitoring Config: configs/train/pretrain-350m.yaml (monitoring section)

monitoring:
  terminal:
    enabled: true
    refresh_rate: 1000              # TUI refresh in ms
    metrics: ["loss", "learning_rate", "gradient_norm"]
    charts:
      - type: "loss_curve"
        metric: "loss"
        window: 100                 # Smoothing window
        show_eta: true

  tracking:
    enabled: true
    backend: "sqlite"               # .entrenar/experiments.db (WAL mode)
    experiment: "albor-pretrain-350m"
    tags:
      model: "albor-350m"
      stage: "pretrain"
      data: "python-code-v2"                 # 139M tokens (v2 dataset)

  system:
    enabled: true
    interval: 5000                  # System metrics every 5s
    metrics: ["gpu_utilization", "memory", "temperature"]

  alerts:
    - condition: "loss > 10"
      action: "stop"
      message: "Loss exploded — Andon stop"
    - condition: "gradient_norm > 100"
      action: "stop"
      message: "Gradient explosion — Andon stop"

6.6.2 What Entrenar Monitors Automatically

6.6.3 Live Monitoring During Training

# Terminal 1 (lambda): Run training
apr train apply --task pretrain --config configs/train/pretrain-350m.yaml

# Terminal 2 (lambda or ssh): Attach live monitor (presentar TUI)
apr monitor ./checkpoints/albor-base-350m/

# Terminal 2 (alternative): JSON output for LLM agents / CI
apr monitor --json ./checkpoints/albor-base-350m/

# Discover all active training runs (reads global SQLite registry)
apr monitor

# List past experiments from SQLite registry
apr runs ls --global

# Show detailed metrics for a specific run
apr runs show <run-id> --global --json

# Browse past experiments from SQLite
apr experiment view --db .entrenar/experiments.db

# Compare loss curves across runs
apr experiment view --db .entrenar/experiments.db \
  --runs albor-pretrain-50m,albor-pretrain-350m \
  --metric loss --chart loss_curve

# One-shot profiler (GPU utilization, per-layer timing)
apr cbtop ./checkpoints/albor-base-350m/latest.safetensors

# Inference latency profiling
apr profile ./checkpoints/albor-base-350m/ --prompt "def fibonacci(n):"

# Stack-level health (from batuta)
batuta stack status

6.6.4 Experiment Lifecycle

Each training run creates two data streams:

Real-time (JSON file IPC) — for live TUI monitoring:

checkpoints/albor-base-350m/
├── training_state.json         # Live metrics (loss, lr, grad_norm, GPU telemetry)
├── checkpoint-step-1000.safetensors
├── checkpoint-step-1000.json   # Checkpoint metadata (epoch, is_best)
├── checkpoint-step-2000.safetensors
├── checkpoint-step-2000.json
├── checkpoint-best.safetensors
└── checkpoint-best.json

Durable (dual SQLite experiment stores) — for post-hoc analysis and comparison:

checkpoints/albor-base-350m/.entrenar/
└── experiments.db              # Local per-experiment store (WAL mode)
    ├── experiments             # Experiment metadata (name, description, config)
    ├── runs                    # Training runs (status, timestamps)
    ├── params                  # Hyperparameters (key/value/type)
    ├── metrics                 # Per-step metrics (loss, lr, tok/s, timestamp)
    ├── artifacts               # Model artifacts (path, size, SHA-256)
    └── span_ids                # Distributed trace integration

~/.entrenar/
└── experiments.db              # Global cross-machine registry (WAL mode)
    └── (same schema)           # All runs across all experiments

PretrainTracker (ALB-055/056) writes to both stores on every log interval. All operations are best-effort — storage failures never block training.

Three consumers, zero contention:

• apr monitor reads training_state.json (atomic write-then-rename) for live dashboards. Multiple monitors attach simultaneously.
• apr runs ls reads ~/.entrenar/experiments.db (global registry) for cross-experiment history. Supports --json for LLM agent consumption.
• apr experiment reads local .entrenar/experiments.db (WAL mode) for per-run metric queries and artifact tracking. Read-only during training — no lock contention with the writer.

6.6.5 Presentar Visualization: Rich Terminal Dashboards

presentar (presentar-terminal) provides ML-specific visualization widgets that entrenar’s TrainingDashboard now composes directly (ALB-057). The dashboard builds a widget tree from Layout::rows() of Border-wrapped section panels, each containing Meter, GpuPanel, Sparkline, or Text widgets. The connection point for historical data is entrenar’s SQLite experiment store (.entrenar/experiments.db).

Live training dashboard (apr monitor — reads training_state.json):

╭─ Albor Pre-Train: albor-base-350m ─── Step 12,847 / 19,073 ──── 67.4% ─╮
│                                                                          │
│  Loss                                          GPU (RTX 4090)            │
│  3.2 ⣀⣀                                       ████████████░░░ 82%       │
│      ⠈⠉⠉⠑⠒⠒⠤⣀                                VRAM: 14.2 / 24.0 GB      │
│               ⠈⠉⠑⠒⠤⣀⣀                        Temp: 72°C                │
│  1.8                  ⠈⠉⠒⠒⣀⣀⣀⣀               Power: 312W               │
│                              ⠉⠉⠉              Tokens/s: 18,432          │
│  0 ──────────────────────────────── 12K                                  │
│                                                                          │
│  Learning Rate              Gradient Norm       ETA: 1d 14h 22m          │
│  ⣿⣿⣿⣷⣶⣶⣤⣤⣀⣀⣀⣀⣀⣀⣀⣀⣀⣀     ▁▁▂▁▁▃▁▂▁▁▁▂▁▁    Throughput: 5.2B / 10B   │
│  3e-4 → 2.1e-4              0.42 (norm)        Checkpoint: step-12000    │
╰──────────────────────────────────────────────────────────────────────────╯

Post-hoc experiment comparison (apr experiment view — reads SQLite):

# Compare loss curves across all pre-training runs
apr experiment view --db .entrenar/experiments.db \
  --runs albor-pretrain-50m,albor-pretrain-350m \
  --metric loss --chart loss_curve

# Hyperparameter comparison table
apr experiment view --db .entrenar/experiments.db \
  --experiment albor-pretrain-350m --params

# Export metrics for external analysis (Parquet for alimentar)
apr experiment export --db .entrenar/experiments.db \
  --run albor-pretrain-350m --format parquet --output ./eval/metrics.parquet

Presentar widgets used by albor:

Two rendering targets, same widgets, same data:

presentar compiles the same widget tree to two targets — terminal and WASM. The dashboard YAML is written once. presentar-terminal renders it via crossterm (works over SSH). presentar renders it via WebGPU in the browser (60fps, GPU-accelerated). Both read from the same data sources.

Both TUI and WASM are first-class deliverables, not stretch goals. The terminal TUI is the primary interface (SSH to lambda/intel). The WASM dashboard is the shareable artifact for model cards and teaching.

6.6.6 No External Dependencies

7. Post-Training Improvement Ladder

Each stage improves the model and exercises a different entrenar / apr capability. Every stage produces a benchmarked checkpoint.

7.1 Stage 1: Pre-Train Base Model

apr train plan configs/train/pretrain-350m.yaml          # Validate + VRAM estimate
apr train apply configs/train/pretrain-350m.yaml --seed 42

Produces: albor-base-350m — raw pre-trained model Exercises: entrenar, trueno (CUDA), alimentar (data streaming) Expected: OPT-350M class on general benchmarks (~48% avg). On HumanEval, target >8% (above random, below CodeGen-350M’s 12.8% due to less training data)

7.2 Stage 2: Knowledge Distillation from Qwen3-Coder-Next

# Plan: check teacher fits in RAM, estimate logit disk usage
apr distill plan configs/train/distill.yaml

# Apply phase 1: Pre-compute teacher logits on intel (300GB RAM, CPU inference)
apr distill apply configs/train/distill.yaml --stage precompute

# Apply phase 2: Distill into student on lambda (4090)
apr distill apply configs/train/distill.yaml --stage train

Produces: albor-distill-350m — distilled model with teacher knowledge Exercises: realizar (teacher inference), apr distill, alimentar (logit storage) Expected: Moderate improvement — absorbs coding patterns from 80B teacher. Estimated +2-7 points on HumanEval via logit-level KD. Note: MoE→dense distillation is uncharted at this scale; the architecture mismatch (DeltaNet+MoE teacher → LLaMA-style dense student) may limit transfer compared to dense→dense distillation (e.g., GPT-3.5→phi-1).

7.3 Stage 3: Instruction Fine-Tuning (LoRA/QLoRA)

apr finetune plan configs/train/finetune-lora.yaml        # Validate LoRA config + VRAM
apr finetune apply configs/train/finetune-lora.yaml

Produces: albor-instruct-350m — instruction-following model Exercises: apr finetune, entrenar LoRA, alimentar (JSONL instruction data) Expected: Better IFEval scores, improved structured output, chat capability.

7.4 Stage 4: Model Merging

apr merge plan \
  --models albor-distill-350m,albor-instruct-350m \
  --method slerp --weight 0.6 \
  --output ./checkpoints/albor-merged/
# Plan checks: architectures compatible, method valid, output size estimate

apr merge apply \
  --models albor-distill-350m,albor-instruct-350m \
  --method slerp --weight 0.6 \
  --output ./checkpoints/albor-merged/

Produces: albor-merged-350m — best-of-all-worlds model Exercises: apr merge (SLERP, TIES, DARE algorithms) Expected: Cherry-picks strengths from each variant. Potentially better than any single model on diverse benchmarks.

7.5 Stage 5: Pruning

apr prune plan \
  --model ./checkpoints/albor-merged-350m/ \
  --method wanda --sparsity 0.5 \
  --output ./checkpoints/albor-pruned/
# Plan checks: model exists, sparsity in [0,1], output size estimate

apr prune apply \
  --model ./checkpoints/albor-merged-350m/ \
  --method wanda --sparsity 0.5 \
  --output ./checkpoints/albor-pruned/

Produces: albor-pruned-175m — half the parameters, similar performance Exercises: apr prune (WANDA, SparseGPT, magnitude, depth pruning) Expected: ~2-5% benchmark degradation at 50% sparsity. WANDA is well-studied at larger scales (7B+) but less validated at 350M where there is less redundancy. Depth pruning to ~18 layers yields ~260M params.

7.6 Stage 6: Quantization

apr quantize plan \
  --model ./checkpoints/albor-merged-350m/ \
  --method q4_k \
  --output ./checkpoints/albor-q4/
# Plan checks: model exists, format valid, output size estimate (~90MB)

apr quantize apply \
  --model ./checkpoints/albor-merged-350m/ \
  --method q4_k \
  --output ./checkpoints/albor-q4/

# Export for broad compatibility
apr export plan --model ./checkpoints/albor-q4/ --format gguf
apr export apply \
  --model ./checkpoints/albor-q4/ \
  --format gguf \
  --output ./release/albor-350m-q4_k.gguf

Produces: albor-q4-350m — 4-bit quantized, ~90MB on disk Exercises: apr quantize, apr export (GGUF, SafeTensors) Expected: <1% benchmark loss from Q4_K quantization. Model runs on any device — phones, Raspberry Pi, browsers (WASM via trueno).

7.7 Benchmark Trajectory

Every stage is benchmarked. The trajectory itself is a key result. Code completion metrics (HumanEval, FIM) are primary; general benchmarks are secondary.

Numbers are estimates. The distillation gain (+2-7 points over base) assumes 500M-2B tokens of teacher logits. This is conservative — published distillation results show larger gains with dense teachers (phi-1 used GPT-3.5, a dense model). Our MoE→dense distillation path is uncharted at 350M scale. The FIM column is removed because there is no standardized FIM benchmark — we will define our own eval and report absolute numbers, not targets. CPU tok/s measured on Xeon at Q4.

8. Evaluation & Benchmarks

8.1 Evaluation Strategy

Leaderboard target: Big Code Models Leaderboard — the standard HuggingFace leaderboard for code generation models. Uses HumanEval (pass@1) and MultiPL-E (18 languages). Currently tracks ~60 models. No sub-1B model has ever appeared on this leaderboard. The smallest entries are 1.0B (DeciCoder-1B at 19.3%, phi-1 at 50.6%, SantaCoder at 18.1%). Albor would be the first sub-1B entry — and the only model trained in Rust.

Secondary: Classic lm-evaluation-harness benchmarks (zero-shot) for general capability comparison against Pythia, OPT, GPT-2.

NOT targeting: Open LLM Leaderboard v2 (IFEval, BBH, MATH Level 5, GPQA, MuSR, MMLU-PRO). These benchmarks were designed for large models — a 350M model scores near random on MATH Level 5 (~0%), GPQA (~25%), and MMLU-PRO (~10%).

Also NOT targeting: EvalPlus Leaderboard (HumanEval+, MBPP+). A secondary submission target if results are strong, but the Big Code leaderboard is the primary scoreboard.

8.2 Benchmark Suite

Python Code Completion Benchmarks (Primary — matches use case)

General Capability Benchmarks (Secondary — validates base model quality)

8.3 Understanding Perplexity

Perplexity is the primary metric for monitoring pre-training progress. It measures how well the model predicts held-out text:

perplexity = e^(cross_entropy_loss)

Intuition: Perplexity is the effective number of tokens the model considers equally likely at each position. A model with perplexity 100 is, on average, choosing between 100 equally probable next tokens. Lower is better — it means the model has learned to concentrate probability mass on the correct tokens.

Scale for albor (vocab_size = 32,768):

Why perplexity, not loss: Cross-entropy loss (ln(perplexity)) compresses the scale. Loss 6.93 vs 6.83 sounds small but corresponds to perplexity 1018 vs 922 — a 10% improvement in prediction quality. Perplexity makes the magnitude of improvements human-readable.

Validation perplexity (val_ppl) is computed on held-out data not seen during training. It detects overfitting: if train loss keeps falling but val_ppl plateaus or rises, the model is memorizing rather than generalizing. The v3 training plateau (val_ppl stuck at ~1000 from step 12K to 28K) was diagnosed via validation perplexity — train loss was still falling slightly, but the model had stopped learning generalizable patterns. Root cause: constant learning rate (ALB-079) and insufficient batch size (ALB-080).

8.4 Competitive Baselines

Python Code Completion Baselines (Primary Competition)

†phi-1-small caveat: phi-1-small was never publicly released — it exists only as an ablation study in “Textbooks Are All You Need” (Gunasekar et al., 2023). The 45% HumanEval claim is self-reported and has never been independently reproduced. We treat it as an aspirational ceiling, not a verified baseline.

The benchmark to beat is CodeGen-350M-mono (same param count, no distillation, no FIM, 12.8% HumanEval). The realistic target for distillation is +2-5 points over the base model. Albor uses a stronger teacher (80B MoE) but faces a significant architecture mismatch (MoE teacher → dense student) and uses a first-generation Rust training stack instead of PyTorch.

Big Code Models Leaderboard — where Albor would land

CodeGen-350M-mono is not on the leaderboard (never submitted). The smallest models currently on the board are 1B-class. If albor-distill hits >15% HumanEval, it would sit just below the 1B tier — at 1/3 the parameter count:

Submission protocol: Run bigcode-evaluation-harness with standard params (top-p=0.95, temperature=0.2, n_samples=50), submit PR to the leaderboard’s community_results/ folder. Results marked as “non-verified” (community).

General Capability Baselines (Secondary)

Note: General capability targets are conservative. Albor is 80% Python code data with a coding teacher — distillation from Qwen3-Coder-Next will not improve general reasoning (ARC-E, HellaSwag). The target is OPT-350M parity, not Pythia-410M. Code benchmarks are the real scoreboard.

8.5 Evaluation Protocol

# Plan: validate model exists, tasks recognized, output writable
apr eval plan \
  --model ./checkpoints/albor-distill-350m/ \
  --tasks humaneval,humaneval_fim,mbpp,ds1000

# Python code completion benchmarks (primary — run after every stage)
apr eval apply \
  --model ./checkpoints/albor-distill-350m/ \
  --tasks humaneval,humaneval_fim,mbpp,ds1000 \
  --output ./eval/python-code-results.json \
  --seed 42

# General capability benchmarks (secondary)
apr eval apply \
  --model ./checkpoints/albor-350m-final/ \
  --tasks arc_easy,hellaswag,piqa,lambada \
  --batch-size 32 \
  --output ./eval/general-results.json \
  --seed 42

# Latency benchmark (critical for code completion use case)
apr bench plan --model ./checkpoints/albor-q4/
apr bench apply \
  --model ./checkpoints/albor-q4/ \
  --prompt "def fibonacci(n):" \
  --max-tokens 128 \
  --device cpu --device cuda \
  --output ./eval/latency-results.json

# Perplexity on held-out Python code
apr eval apply \
  --model ./checkpoints/albor-350m-final/ \
  --perplexity \
  --data ./data/eval/held-out-python.parquet

# ── Big Code Leaderboard submission eval ──
# Must use bigcode-evaluation-harness with standard params for comparability
# This runs OUTSIDE the sovereign stack (Python, not Rust) — it is the
# leaderboard's own eval tool, not ours. Our apr eval results are the
# primary record; this is for leaderboard submission only.
#
# bigcode-evaluation-harness \
#   --model ./release/albor-350m.safetensors \
#   --tasks humaneval,multiple-py \
#   --temperature 0.2 --top_p 0.95 \
#   --n_samples 50 --max_length_generation 512 \
#   --output ./eval/bigcode-leaderboard/

8.6 Continuous Evaluation During Training

The intel box runs eval on the latest checkpoint concurrently with training:

# On intel (300GB RAM), polling for new checkpoints
apr eval apply \
  --model ./checkpoints/latest/ \
  --tasks arc_easy,hellaswag \
  --batch-size 16 \
  --output ./eval/step-$(cat ./checkpoints/latest/step.txt).json

Gap ALB-006: Verify apr eval plan/apply supports these benchmark tasks natively. FIXED: apr eval supports perplexity and classification eval.

Gap ALB-037 (FIXED): apr eval previously ignored loaded weights during inference. Now fixed — realizar run loads trained SafeTensors checkpoints and generates from learned weights. Verified end-to-end with 350M test checkpoint (218 tensors loaded, tokens generated). scripts/eval-perplexity.py provides independent pure-Python perplexity evaluation.

Gap ALB-038 (FIXED): entrenar previously saved initialization weights instead of trained weights due to broken autograd gradient flow. Root cause: RMSNorm::forward_batched() created tensors with no backward op, and MultiHeadAttention::forward() broke Q/K/V gradient chain. Fixed in entrenar@91ba9da (RMSNorm backward) and entrenar@1ede409 (attention backward). All 20 model parameters now receive gradients during training. See GitHub #36.

Gap ALB-059 (FIXED): GEMM backward constructor args n/k swapped in entrenar — baked wrong compile-time stride constants into PTX. Output rows overflowed into optimizer state buffers, causing NaN in AdamW. The 50-step test model trained with this bug had loss 10.39→6.07; after the fix, loss improved to 10.39→5.92. All evaluation results should use the post-fix checkpoint (entrenar@846ae0c). Additionally, all optimizer m/v buffers are now zero-initialized (cuMemAlloc returns uninitialized VRAM).

Gap ALB-060 (CONFIG FIXED): The original “full” 350M training run completed only 43/5000 steps because epochs: 1 with grad_accum: 128 exhausted the 22K-sequence dataset. Fix: C-TRAINCFG-001 contract + v2 config (pretrain-350m-v2.yaml) with expanded 68K-sequence dataset, epochs: 1 (steps_per_epoch = 16994 >= 5000), gradient_accumulation: 1 (ALB-066). The v2 training run (ALB-063) reached step ~1183/5000, loss 10.4→6.9 (clear convergence), then stopped. The checkpoints/albor-base-350m-v2/ checkpoint has partially trained weights. Full evaluation awaits training completion.

8.7 Local Evaluation Infrastructure

The following scripts provide model evaluation independently of apr eval:

# Validate checkpoint integrity (fast, detects ALB-038)
python scripts/eval-perplexity.py checkpoints/albor-base-350m/ --validate-checkpoint

# Validate all canonical solutions (no model needed)
python scripts/eval-code.py configs/eval/python-intermediate.jsonl --validate-only
python scripts/eval-code.py configs/eval/humaneval-subset.jsonl --validate-only

# Full evaluation suite (orchestrates all steps)
bash scripts/run-eval-suite.sh checkpoints/albor-base-350m/

# Perplexity on pre-tokenized validation data
python scripts/eval-perplexity.py checkpoints/albor-base-350m/ \
    --data data/pretokenized-2048/val/val.parquet \
    --max-sequences 100 --seq-len 2048 --threshold 30

# Evaluate via apr serve API (ALB-037 FIXED — realizar loads trained checkpoints)
python scripts/eval-code.py configs/eval/humaneval-subset.jsonl \
    --api http://localhost:8080 --samples 10

# Training convergence validation (FALSIFY-ALBOR-001)
python scripts/validate-training-convergence.py \
    checkpoints/albor-base-350m/training.log

# Convert entrenar checkpoint format for realizar
python scripts/convert-checkpoint.py checkpoints/albor-base-350m/ \
    --config configs/train/pretrain-350m.yaml

Benchmark datasets:

• configs/eval/python-intermediate.jsonl — 15 intermediate Python problems
• configs/eval/humaneval-subset.jsonl — 20 HumanEval-format problems

8.8 Weight Convention & Checkpoint Format

entrenar stores linear layer weights as [in_features, out_features] in row-major (C) order, and computes forward pass as x @ W (no transpose). This differs from the HuggingFace convention of [out_features, in_features] with x @ W.T.

The convert-checkpoint.py script handles the conversion:

1. Reads 1D flat tensors from entrenar SafeTensors
2. Reshapes as [in, out] (entrenar convention)
3. Transposes to [out, in] (HuggingFace/realizar convention)
4. Writes new SafeTensors with proper 2D shapes

Embeddings (model.embed_tokens.weight) are stored as [vocab, hidden] in both conventions (indexed by token ID for row lookup).

9. Distributed Training Architecture

9.1 Machine Roles (Revised)

With 300 GB RAM on the intel box, the architecture is asymmetric:

9.2 Distillation Split (Primary Distributed Architecture)

The natural multi-machine split is teacher on intel, student on lambda:

┌───────────────────────────────┐                          ┌───────────────────────────┐
│  intel (300 GB RAM)           │    pre-computed logits    │  lambda (RTX 4090)        │
│                               │    as sharded Parquet     │                           │
│  Qwen3-Coder-Next 80B fp16   │ ────────────────────────► │  albor-350M student       │
│  Full model in CPU RAM        │    (rsync / NFS)          │  KD loss + CE loss        │
│  realizar CPU inference       │                           │  Full GPU speed training  │
│  ~5-15 tok/s                  │                           │                           │
│                               │ ◄──── checkpoints ─────  │  apr distill apply    │
│  Concurrent eval runner       │    (rsync / NFS)          │                           │
└───────────────────────────────┘                           └───────────────────────────┘

This requires no gradient sync, no ring all-reduce, no distributed training framework for the distillation stage. The teacher pre-computes logits offline; the student trains at full GPU speed against stored logits. Simple and effective.

9.3 Entrenar Native DDP (Complete)

entrenar has full distributed data parallelism infrastructure (entrenar#133), superseding the repartir approach:

Implemented (all wired end-to-end):

• Wire protocol v2: TCP-based message framing with BlockGradientPayload, AveragedBlockGradient, NonBlockGradientPayload, AveragedNonBlockGradient
• GradientServer: Coordinator that collects gradients from N workers, averages them (per-block AllReduce), and broadcasts averaged gradients back
• WorkerClient: Worker-side TCP client that sends/receives gradient payloads
• PerBlockGradientAccumulator: CPU-side gradient accumulator for AllReduce (same one used by ALB-066 single-GPU gradient accumulation)
• RingAllReduce: Ring-based averaging for N workers
• DistributedCudaTrainer: train_batch() → forward+backward → per-block AllReduce → optimizer step. Wraps CudaTransformerTrainer with distributed comm
• train_loop_cuda_distributed(): Full training loop with data sharding by rank, coordinator thread auto-spawn (rank 0), worker connection, epoch iteration
• spawn_coordinator_thread(): Background thread running GradientServer for rank 0 process
• CLI flags: --distributed --world-size N --rank R inject distributed config into YAML at runtime
• 11 integration tests: C-DDP-001 weight consistency via BLAKE3, 4-worker ring AllReduce, per-block reverse-order AllReduce

Architecture:

Process 0 (rank=0):                     Process 1 (rank=1):
  GradientServer (bg thread)
  DistributedCudaTrainer                  DistributedCudaTrainer
    └─ CudaTransformerTrainer (GPU 0)       └─ CudaTransformerTrainer (GPU 1)
    └─ WorkerClient → TCP ─────────────────── WorkerClient → TCP

9.4 Original Repartir Gaps (Stretch)

The original plan for distributed training via a standalone repartir crate is now partially superseded by entrenar’s native DDP, but some gaps remain relevant for cross-vendor GPU support:

Gap ALB-002: Ring all-reduce (now partially implemented in entrenar itself). Gap ALB-004: Unified CUDA + wgpu backend dispatch in entrenar. Gap ALB-005: trueno wgpu backward pass (gradient WGSL shaders).

The distillation architecture (Section 9.2) achieves multi-machine utilization without any of these.

9.5 W5700X Role

The W5700X GPUs (2x 8GB each) can assist with:

• Eval inference: Run benchmarks on latest checkpoint via wgpu/Vulkan
• Partial KV cache offload: Assist CPU-based teacher inference
• Future: Participate in gradient-parallel training once ALB-005 is resolved

10. Pipeline Orchestration (apr pipeline + forjar DAG)

10.1 Architecture: One Manifest, One DAG

The entire albor pipeline — from bare metal to published model — lives in a single YAML manifest: configs/pipeline/albor.yaml. Forjar’s DAG engine resolves dependencies, tracks state, and dispatches steps across machines. apr pipeline wraps forjar, so the user never calls forjar directly.

apr pipeline plan configs/pipeline/albor.yaml    # Show full DAG, estimate everything
apr pipeline apply configs/pipeline/albor.yaml   # Execute (resumable)
apr pipeline status                              # Show what's converged/pending/failed
apr pipeline drift                               # Detect unauthorized state changes

How it works:

                     configs/pipeline/albor.yaml
                              │
                    apr pipeline plan/apply
                              │
                     forjar DAG engine
                    (Kahn's toposort)
                              │
         ┌────────────┬───────┴───────┬────────────┐
         │            │               │            │
    infra resources   │          task resources    │
    (package, gpu,    │          (run apr cmds,    │
     file, mount,     │           track output)    │
     model)           │               │            │
         │            │               │            │
    forjar native     │     apr train apply        │
    convergence       │     apr distill apply      │
                      │     apr eval apply         │
                      │     apr publish apply      │
                      │               │            │
                 state/lambda/     state/intel/
                 state.lock.yaml   state.lock.yaml

Key properties:

• Resumable: BLAKE3 hashes per resource. Re-run skips converged steps.
• Multi-machine: Infra + tasks dispatch to lambda or intel via SSH.
• Plan/apply: apr pipeline plan shows the full DAG with estimates before committing any resources. Exit 0 if valid, exit 1 with diagnostics.
• Idempotent: Same manifest, same state → zero changes (all NoOp).
• bashrs linted: All shell fragments in task command: fields are validated by bashrs (Rash v6.65) at plan time. No unvalidated shell reaches execution. bashrs is KING of linting — bashrs make lint validates Makefiles, bashrs lint validates shell scripts, bashrs classify classifies safety.

Dual orchestration:

• forjar manifest (configs/pipeline/albor.yaml): Infrastructure provisioning (GPU drivers, packages, directories, mounts, teacher model download). Blocked on type: task (ALB-027) for ML steps.
• batuta playbook (configs/pipeline/albor-playbook.yaml): ML pipeline orchestration (data prep, train, distill, finetune, merge, prune, quantize, eval, publish). 19-stage deterministic DAG with BLAKE3 caching. Validates successfully.

10.2 Pipeline Manifest: configs/pipeline/albor.yaml

version: "1.0"
name: albor-training-pipeline
description: "Sovereign Python code completion model — full pipeline"

machines:
  lambda:
    hostname: lambda
    addr: 127.0.0.1
    user: noah
    arch: x86_64
    roles: [gpu-train, student]

  intel:
    hostname: intel
    addr: intel
    user: noah
    ssh_key: ~/.ssh/id_ed25519
    arch: x86_64
    roles: [teacher-inference, data-pipeline, eval, checkpoint-backup]

resources:
  # ═══════════════════════════════════════════════════════════
  # INFRASTRUCTURE (forjar native resources)
  # ═══════════════════════════════════════════════════════════

  cuda-driver:
    type: gpu
    machine: lambda
    gpu_backend: nvidia
    driver_version: "550"
    cuda_version: "12.4"
    persistence_mode: true
    compute_mode: exclusive_process

  vulkan-driver:
    type: package
    machine: intel
    provider: apt
    state: present
    packages: [mesa-vulkan-drivers, vulkan-tools, libvulkan-dev]

  data-dir:
    type: file
    machine: [lambda, intel]
    path: /data/albor
    state: directory
    mode: "0755"

  teacher-model:
    type: model
    machine: intel
    name: Qwen/Qwen3-Coder-Next
    state: present
    cache_dir: /data/albor/models/teacher
    depends_on: [data-dir]

  checkpoint-share:
    type: mount
    machine: intel
    source: "lambda:/data/albor/checkpoints"
    path: /data/albor/checkpoints
    fstype: nfs
    options: "rw,sync,no_subtree_check"
    depends_on: [data-dir]

  logit-share:
    type: mount
    machine: lambda
    source: "intel:/data/albor/teacher-logits"
    path: /data/albor/teacher-logits
    fstype: nfs
    options: "ro,sync"
    depends_on: [data-dir]

  # ═══════════════════════════════════════════════════════════
  # DATA PIPELINE (task resources — call apr subcommands)
  # ═══════════════════════════════════════════════════════════

  ingest-local:
    type: task
    machine: lambda
    command: >
      alimentar import local ../depyler/examples/ ../depyler/tdd-book/tests/
        --lang python --output ./data/local/depyler.parquet &&
      alimentar import local ../hf-ground-truth-corpus/
        --lang python --output ./data/local/hf-gtc.parquet &&
      alimentar import local ../jax-ground-truth-corpus/
        --lang python --output ./data/local/jax-gtc.parquet &&
      alimentar import local ../vllm-ground-truth-corpus/
        --lang python --output ./data/local/vllm-gtc.parquet
    output_artifacts: ["./data/local/*.parquet"]
    depends_on: [data-dir]

  ingest-external:
    type: task
    machine: lambda
    command: >
      alimentar import hf bigcode/starcoderdata --lang python
        --output ./data/starcoder-python/ &&
      alimentar import hf HuggingFaceFW/fineweb-edu
        --output ./data/fineweb-edu/
    output_artifacts: ["./data/starcoder-python/", "./data/fineweb-edu/"]
    depends_on: [data-dir]

  data-mix:
    type: task
    machine: lambda
    command: >
      alimentar quality check ./data/ --profile ml-training &&
      alimentar mix
        --input ./data/local/depyler.parquet --weight 0.025 --upsample 10
        --input ./data/local/hf-gtc.parquet --weight 0.025 --upsample 10
        --input ./data/local/jax-gtc.parquet --weight 0.025 --upsample 10
        --input ./data/local/vllm-gtc.parquet --weight 0.025 --upsample 10
        --input ./data/starcoder-python/ --weight 0.40
        --input ./data/fineweb-edu/ --weight 0.20
        --input ./data/processed/python-docs.parquet --weight 0.10
        --output ./data/mixed/ --seed 42 --shuffle
    output_artifacts: ["./data/mixed/"]
    depends_on: [ingest-local, ingest-external]

  tokenize:
    type: task
    machine: lambda
    command: >
      apr tokenize plan --input ./data/mixed/*.parquet --vocab-size 32768
        --output ./models/albor-tokenizer/ &&
      apr tokenize apply --input ./data/mixed/*.parquet --vocab-size 32768
        --output ./models/albor-tokenizer/ --seed 42 &&
      apr tokenize apply --tokenizer ./models/albor-tokenizer/
        --input ./data/mixed/*.parquet --output ./data/tokenized/
        --max-seq-len 2048
    output_artifacts: ["./models/albor-tokenizer/", "./data/tokenized/"]
    depends_on: [data-mix]

  # ═══════════════════════════════════════════════════════════
  # TRAINING (task resources — long-running, checkpoint-aware)
  # ═══════════════════════════════════════════════════════════

  train-50m:
    type: task
    machine: lambda
    command: >
      apr train plan configs/train/pretrain-50m.yaml &&
      apr train apply configs/train/pretrain-50m.yaml --seed 42
    output_artifacts: ["./checkpoints/albor-base-50m/"]
    completion_check: "test -f ./checkpoints/albor-base-50m/checkpoint-best.safetensors"
    depends_on: [tokenize, cuda-driver]

  train-350m:
    type: task
    machine: lambda
    command: >
      apr train plan configs/train/pretrain-350m.yaml &&
      apr train apply configs/train/pretrain-350m.yaml --seed 42
    output_artifacts: ["./checkpoints/albor-base-350m/"]
    completion_check: "test -f ./checkpoints/albor-base-350m/checkpoint-best.safetensors"
    depends_on: [train-50m]

  # ═══════════════════════════════════════════════════════════
  # DISTILLATION (cross-machine: intel produces logits, lambda trains)
  # ═══════════════════════════════════════════════════════════

  distill-logits:
    type: task
    machine: intel
    command: >
      apr distill plan configs/train/distill.yaml &&
      apr distill apply configs/train/distill.yaml --stage precompute
    output_artifacts: ["./data/teacher-logits/"]
    completion_check: "test -d ./data/teacher-logits/ && ls ./data/teacher-logits/*.parquet"
    depends_on: [train-350m, teacher-model, logit-share]

  distill:
    type: task
    machine: lambda
    command: >
      apr distill apply configs/train/distill.yaml --stage train --seed 42
    output_artifacts: ["./checkpoints/albor-distill/"]
    completion_check: "test -f ./checkpoints/albor-distill/checkpoint-best.safetensors"
    depends_on: [distill-logits]

  # ═══════════════════════════════════════════════════════════
  # POST-TRAINING LADDER (sequential, each depends on previous)
  # ═══════════════════════════════════════════════════════════

  finetune:
    type: task
    machine: lambda
    command: >
      apr finetune plan configs/train/finetune-lora.yaml &&
      apr finetune apply configs/train/finetune-lora.yaml
    output_artifacts: ["./checkpoints/albor-instruct/"]
    depends_on: [distill]

  merge:
    type: task
    machine: lambda
    command: >
      apr merge plan --models albor-distill-350m,albor-instruct-350m
        --method slerp --weight 0.6 --output ./checkpoints/albor-merged/ &&
      apr merge apply --models albor-distill-350m,albor-instruct-350m
        --method slerp --weight 0.6 --output ./checkpoints/albor-merged/
    output_artifacts: ["./checkpoints/albor-merged/"]
    depends_on: [finetune]

  prune:
    type: task
    machine: lambda
    command: >
      apr prune plan --model ./checkpoints/albor-merged-350m/
        --method wanda --sparsity 0.5 --output ./checkpoints/albor-pruned/ &&
      apr prune apply --model ./checkpoints/albor-merged-350m/
        --method wanda --sparsity 0.5 --output ./checkpoints/albor-pruned/
    output_artifacts: ["./checkpoints/albor-pruned/"]
    depends_on: [merge]

  quantize:
    type: task
    machine: lambda
    command: >
      apr quantize plan --model ./checkpoints/albor-merged-350m/
        --method q4_k --output ./checkpoints/albor-q4/ &&
      apr quantize apply --model ./checkpoints/albor-merged-350m/
        --method q4_k --output ./checkpoints/albor-q4/
    output_artifacts: ["./checkpoints/albor-q4/"]
    depends_on: [merge]

  # ═══════════════════════════════════════════════════════════
  # EVALUATION (can run on intel concurrently with training)
  # ═══════════════════════════════════════════════════════════

  eval-code:
    type: task
    machine: lambda
    command: >
      apr eval plan --model ./checkpoints/albor-merged-350m/
        --tasks humaneval,humaneval_fim,mbpp,ds1000 &&
      apr eval apply --model ./checkpoints/albor-merged-350m/
        --tasks humaneval,humaneval_fim,mbpp,ds1000
        --output ./eval/python-code-results.json --seed 42
    output_artifacts: ["./eval/python-code-results.json"]
    depends_on: [merge]

  eval-general:
    type: task
    machine: intel
    command: >
      apr eval apply --model ./checkpoints/albor-merged-350m/
        --tasks arc_easy,hellaswag,piqa,lambada
        --output ./eval/general-results.json --seed 42
    output_artifacts: ["./eval/general-results.json"]
    depends_on: [merge, checkpoint-share]

  # ═══════════════════════════════════════════════════════════
  # RELEASE
  # ═══════════════════════════════════════════════════════════

  export:
    type: task
    machine: lambda
    command: >
      apr export plan --model ./checkpoints/albor-q4/ --format gguf &&
      apr export apply --model ./checkpoints/albor-q4/ --format gguf
        --output ./release/albor-350m-q4_k.gguf &&
      apr export apply --model ./checkpoints/albor-merged-350m/
        --format safetensors
        --output ./release/albor-350m.safetensors
    output_artifacts: ["./release/"]
    depends_on: [quantize, eval-code]

  publish:
    type: task
    machine: lambda
    command: >
      apr publish plan --model ./release/ --hub paiml/albor-350m &&
      apr publish apply --model ./release/ --hub paiml/albor-350m
    depends_on: [export, eval-general]

policy:
  failure: stop_on_first
  parallel_machines: true
  retry: 2
  bashrs_lint: true            # Validate all task command: fields via bashrs

10.3 Pipeline Workflow

# Show full DAG with time/resource estimates (no side effects)
apr pipeline plan configs/pipeline/albor.yaml

# Execute everything (resumable — skips converged steps)
apr pipeline apply configs/pipeline/albor.yaml

# Check what's done, what's pending, what failed
apr pipeline status

# Detect unauthorized changes to converged resources
apr pipeline drift

# Re-run only failed steps (everything else is NoOp)
apr pipeline apply configs/pipeline/albor.yaml

# Force re-run a specific resource and its dependents
apr pipeline apply configs/pipeline/albor.yaml --target train-350m --force

10.4 The task Resource Type (ALB-027)

The task resource is what makes forjar a pipeline orchestrator, not just an infrastructure tool. It runs an arbitrary command, tracks completion, and hashes output artifacts for idempotency.

Idempotency for ML tasks: A task resource is considered converged when:

1. The command exited 0 on a previous run, AND
2. The BLAKE3 hash of output_artifacts matches the lock file, AND
3. The completion_check (if set) passes

If any of these fail, the task is re-run. For training jobs that crashed mid-run, the command itself includes --resume logic (e.g., apr train apply auto-detects and resumes from the latest checkpoint).

10.5 Why Not Makefile / Shell Scripts

The Makefile and shell scripts are eliminated. One manifest. One DAG. One tool.

11. Gap Register

Every gap discovered during development is tracked here. Each gap maps to a specific upstream component, a GitHub issue, and a clear acceptance criterion.

Lifecycle: Gap discovered → GitHub issue filed → implemented upstream → wired into apr → dogfooded in albor pipeline → FALSIFY/pmat verified → closed.

11.1 Critical Path Gaps (Block the Improvement Ladder)

11.2 Distributed Training Gaps (Stretch / Future)

11.3 Quality & Verification Gaps

Contract coverage report (pv coverage contracts): 8 contracts, 31 equations, 51 obligations, 34 falsification tests, 10 Kani harnesses, 100% obligation coverage. All contracts at impl=0/N — waiting for upstream bindings.

11.4 Dogfooding-Discovered Gaps

| ALB-061 | #43 | albor docs | Monolithic spec stale — diverges from mdBook chapters | Medium | FIXED | scripts/generate-spec.sh regenerates docs/specifications/albor-llm-spec.md from mdBook chapters. make spec target added. | | ALB-062 | #44 | albor docs | Stale spec chapters — §3 VRAM, §15/18 blockers, §16 repro, model card, intro | Medium | FIXED | All chapters updated to match reality: VRAM budget, ALB-025/037 no longer blockers, v2 pipeline in §16, ALB-060 context in model card and introduction. | | ALB-063 | #45 | albor training | Retrain 350M with v2 config (corrected epochs + expanded data) | Critical | IN PROGRESS | ALB-069→072 all fixed. Training running: PID 1775202, ~4.4s/step (934 tok/s), save_interval=250, 5000 steps, ~11.8 GB VRAM. Loss 10.40→7.13 (step 169)→6.77 (step 338). Step 250 eval: val_loss=6.92, val_ppl=1008. Step 500 checkpoint verified OK (1520 MB). gnorm stable 2-9 range. | | ALB-064 | #46 | albor / entrenar | Training process dies silently — no crash detection, no watchdog, no recovery | Critical | FIXED | scripts/train-guard.sh: crash-resilient supervisor with exit code classification, GPU state capture, structured JSON crash reports, exponential backoff restart, heartbeat monitoring, pre-flight GPU health checks. Auto-diagnostic mode: detects async CUDA crash pattern, enables CUDA_LAUNCH_BLOCKING=1 on restart. Five Whys: CUDA driver crash → SIGABRT/SIGSEGV → bypasses Rust panic handler → no stderr output → no diagnosis. Root cause: ALB-065. | | ALB-065 | #47 | entrenar / trueno | Missing stream.synchronize() before D2H gradient transfers — async CUDA crash | Critical | FIXED | compute_workspace_clip_scale() and compute_clip_scale() call cuMemcpyDtoH without synchronizing the non-blocking CUDA stream. cuMemcpyDtoH only synchronizes with the default stream, but trueno creates streams with CU_STREAM_NON_BLOCKING. Result: backward kernels not finished when gradient buffers are read → garbage clip scale → NaN/crash. Fix: stream.synchronize() at 3 locations before D2H transfers (entrenar@d3a3d26). |

| ALB-066 | #48 | albor config | gradient_accumulation: 128 makes training take 68.8 days on single GPU | Critical | FIXED | CudaTransformerTrainer does per-sequence optimizer updates (per-block interleaved backward+optimize). gradient_accumulation just increases sequences per “step” without changing update granularity. Fix: reduced 128→16→1, epochs from 38→5→1. New estimate: ~11.7h at 480 tok/s. | | ALB-067 | #49 | entrenar / trueno | Per-block weight gradient clipping CPU bottleneck — 864 D2H transfers/step | High | FIXED (via ALB-078) | compute_workspace_clip_scale downloaded 9 buffers × 24 blocks × 4 seqs = 864 D2H transfers/step. Workaround: disabled per-block clipping (entrenar@eaadbc6). Proper fix: ALB-078 fused GPU clip pipeline (zero D2H, zero sync). grad_clip: 1.0 re-enabled in v3 config. | | ALB-068 | #50 | entrenar | save_interval dead code — no intermediate checkpoint saving during CUDA training | Critical | FIXED | save_interval read from config, validated, but never used in train_loop_cuda(). Checkpoints only saved at training completion. 24h crash = total loss. Fix: manual batch loop with trainer.save() at save_interval boundaries (entrenar@d8dfab7). | | ALB-069 | #51 | trueno | PTX selp_f32 argument order bug in fused cross-entropy kernels — training produces loss=0.0 | Critical | FIXED | selp_f32(pred, true_val, false_val) called as selp_f32(grad_target, grad_nontarget, is_target) — f32 values in pred slot, predicate in false_val slot. PTX JIT fails: “Arguments mismatch for instruction ‘selp’”. Same class as ALB-059 (constructor arg ordering). Fix: selp_f32(is_target, grad_target, grad_nontarget) at both call sites (trueno@10bec89, trueno#156). | | ALB-070 | #52 | entrenar / albor config | save_interval YAML field ignored — bridge reads checkpoint.save_every, default=1 causes eval every step | Critical | FIXED | YAML bridge reads training.checkpoint.save_every, not training.save_interval. Default=1 → validation eval runs every step → eval_batch() crashes on long sequences (missing max_seq_len truncation). Two fixes: (1) YAML config moved to checkpoint.save_every: 25 (2) eval_batch() now truncates to max_seq_len (entrenar@5c4c2d8). Same class as ALB-060 (config field mismatch). | | ALB-071 | #53 | entrenar | Embed gradient clipping disabled when grad_clip=None — NaN weights, loss=0.0 by step ~100 | Critical | FIXED | C-EMBED-GRAD-001 was gated behind if let Some(max_norm) = max_grad_norm. ALB-067 disabled grad_clip → embed activation gradients unclipped → CPU AdamW overflow → 304K NaN in embeddings, block weights ALL NaN. Fix: always clip with unwrap_or(1.0) + always compute LM head grad norm for observability (entrenar@d07d67d). Same class as ALB-044. | | ALB-072 | #54 | entrenar | fp16 loss scaling causes NaN in early layers — gradient overflow in f32 backward | Critical | FIXED | fp16 GradScaler (scale=65536) multiplied into fused CE kernel’s loss_scale. All backward uses f32 GpuBuffers — no fp16 underflow risk, but 65536x scaling caused activation gradient overflow by layers 0-1. Five Whys: loss=0.0 → NaN blocks 0-1 → first optimizer step NaN → FP32 works/FP16 doesn’t → unnecessary 65536x scaling. Fix: exclude grad_scaler.scale() from loss_scale (entrenar@44d3e74). gnorm now matches FP32 baseline (2.29). | | ALB-073 | #55 | trueno | fused_cross_entropy PTX selp argument mismatch — JIT compilation failure | High | FIXED | Same class as ALB-069. selp_f32(true_val, false_val, pred) instead of (pred, true_val, false_val) in fused cross-entropy kernel. Training fell back to non-fused path. Fix: trueno@10bec89. | | ALB-074 | #56 | entrenar | Buffer overflow — 2048-token seq hits 1024-sized GPU buffer during eval | Critical | FIXED | Stale binary missed ALB-070 eval truncation fix. 2048-token pretokenized sequence passed to eval_single_sequence without max_seq_len truncation → slice overflow at cuda_trainer.rs:711 (2096128 > 1048576). Crashed at step 1183. Fix: binary rebuild with entrenar@5c4c2d8. |

11.5 Performance Optimization Gaps

| ALB-078 | trueno #171, entrenar #240 | trueno / entrenar | Fused GPU gradient clipping — eliminate 26 stream syncs/step | High | IMPLEMENTED | Per-block clip calls stream.synchronize() + D2H 24×/step. New kernels: ClipScaleReduceKernel (single-CTA norm+clip_scale on GPU), GradientClipGpuScaleKernel (element-wise clip reading scale from GPU memory). Pipeline: 9× squared_sum_launch_into → 1× clip_scale_reduce → 9× gradient_clip_gpu_scale. Zero sync, zero D2H. IEEE 754 handles zero-norm (div→+inf, min→1.0). Compiles, awaiting dogfood. Expected: ~20% step time reduction. |

11.6 Training Quality Gaps

11.7 Data Pipeline Gaps

Gaps are added as they are discovered during implementation and dogfooding.

12. Provable Quality & Design by Contract

Every computational kernel used in Albor must have a provable-contracts YAML specification with Popperian falsification tests, property-based probar tests, and Kani bounded model checking harnesses. This is not optional — it is a first-class deliverable alongside the model.

12.1 Verification Ladder

Four levels of assurance, from cheapest to most rigorous:

Level 4: Kani bounded model check    ─── PROOF (exhaustive for inputs ≤ N)
Level 3: probar property tests       ─── HIGH CONFIDENCE (10,000+ random inputs)
Level 2: Falsification tests         ─── TARGETED (specific edge cases)
Level 1: Type system                 ─── BY CONSTRUCTION (Rust compiler)
Level 0: Code review                 ─── HUMAN (necessary but insufficient)

Requirement: Every kernel reaches at least Level 3. Critical kernels (softmax, attention, cross-entropy, KD loss) reach Level 4.

12.2 Contract Registry for Albor

Albor requires contracts for every kernel in the training + post-training pipeline. Many already exist in provable-contracts; new ones must be written.

Existing Contracts (bind to aprender implementations)

New Contracts Required for Albor (ALB-013 through ALB-017)

12.3 Contract Workflow for Each Kernel

# 1. Write or validate YAML contract
pv validate contracts/knowledge-distillation-kernel-v1.yaml

# 2. Generate trait stubs + failing tests
pv scaffold contracts/knowledge-distillation-kernel-v1.yaml

# 3. Generate property-based tests (wired to actual aprender code)
pv probar contracts/knowledge-distillation-kernel-v1.yaml \
  --binding contracts/aprender/binding.yaml

# 4. Generate Kani bounded model checking harnesses
pv kani contracts/knowledge-distillation-kernel-v1.yaml

# 5. Run falsification sweep
pv audit contracts/knowledge-distillation-kernel-v1.yaml \
  --binding contracts/aprender/binding.yaml

# 6. Verify full contract status
pv status contracts/knowledge-distillation-kernel-v1.yaml

12.4 Falsification Tests: Albor-Specific

Every claim in this specification must be falsifiable. Below are the concrete falsification tests for Albor’s key properties.

Training Correctness

# FALSIFY-ALBOR-001: Loss decreases monotonically (smoothed)
- id: FALSIFY-ALBOR-001
  rule: "Training convergence"
  prediction: "EMA(loss, window=100) is monotonically decreasing after warmup"
  test: "Load training log, compute EMA, assert no sustained increase >5% over 500 steps"
  if_fails: "Learning rate too high, data corruption, or gradient computation bug"

# FALSIFY-ALBOR-002: Gradient norms are bounded
- id: FALSIFY-ALBOR-002
  rule: "Training stability"
  prediction: "Global gradient norm < 10.0 after clipping for all steps"
  test: "Parse training log, assert max gradient norm across all steps"
  if_fails: "Gradient clipping not applied, loss spike, or NaN propagation"

# FALSIFY-ALBOR-003: Checkpoint determinism
- id: FALSIFY-ALBOR-003
  rule: "Reproducibility"
  prediction: "Two runs with seed=42 produce identical checkpoints at step 1000"
  test: "Train twice, BLAKE3 hash both checkpoints, assert equality"
  if_fails: "Non-deterministic operation (async GPU, HashMap ordering, etc.)"

Distillation Correctness

# FALSIFY-ALBOR-004: KL divergence is non-negative
- id: FALSIFY-ALBOR-004
  rule: "KD loss validity"
  prediction: "KL(teacher || student) >= 0 for all batches"
  test: "proptest with 10000 random logit pairs, assert KL >= -1e-7"
  if_fails: "Log-domain computation error or softmax numerical instability"

# FALSIFY-ALBOR-005: Distillation improves over base
- id: FALSIFY-ALBOR-005
  rule: "Distillation value"
  prediction: "albor-distill avg benchmark > albor-base avg benchmark"
  test: "Run full eval suite on both, paired t-test with p < 0.05"
  if_fails: "Teacher logits corrupted, temperature too high/low, or alpha miscalibrated"

# FALSIFY-ALBOR-006: Teacher logit integrity
- id: FALSIFY-ALBOR-006
  rule: "Data pipeline integrity"
  prediction: "Pre-computed teacher logits match live teacher inference within 1e-4"
  test: "Sample 100 batches, run live teacher inference, compare against stored logits"
  if_fails: "Serialization precision loss, wrong batch ordering, or teacher model mismatch"

Post-Training Invariants

# FALSIFY-ALBOR-007: Merge interpolation bound
- id: FALSIFY-ALBOR-007
  rule: "SLERP correctness"
  prediction: "‖SLERP(w1, w2, t)‖ ≈ ‖w1‖ for t ∈ [0,1] (unit sphere)"
  test: "proptest with 10000 random weight pairs and t values"
  if_fails: "SLERP implementation uses LERP instead, or normalization missing"

# FALSIFY-ALBOR-008: Pruning sparsity guarantee
- id: FALSIFY-ALBOR-008
  rule: "WANDA correctness"
  prediction: "Exactly 50% of weights are zero after prune --sparsity 0.5"
  test: "Count zero weights, assert within ±0.1% of target sparsity"
  if_fails: "Pruning threshold computation error or layer exclusion bug"

# FALSIFY-ALBOR-009: Quantization round-trip
- id: FALSIFY-ALBOR-009
  rule: "Q4 fidelity"
  prediction: "Perplexity(Q4 model) < 1.05 × Perplexity(fp16 model)"
  test: "Evaluate both on held-out set, assert ratio < 1.05"
  if_fails: "Quantization calibration data insufficient or block size wrong"

12.5 Brick Profiling Architecture

Training a 350M model on a single 4090 is a systems engineering problem, not a scaling problem. Every watt of GPU silicon must be accounted for. The architecture achieves this by treating each component as a brick — a self-contained unit with measurable inputs, outputs, and a provable contract.

12.5.1 Three Granularities of Profiling

Per-kernel. Every CUDA kernel (gemm_forward, silu_backward, rms_norm_forward, batched_transpose_forward, etc.) is individually measurable via compute-sanitizer, nsys, or nvprof. When a kernel misbehaves, the brick boundary isolates the failure to a single function with known input/output shapes. The contract for each kernel specifies buffer size invariants that can be checked statically.

Per-block. CudaTransformerBlock encapsulates one transformer layer’s forward, backward, and optimizer step as a single GPU-resident unit. Diagnostic sampling after backward (downloading 1K elements from each gradient buffer) immediately distinguishes “math is wrong” (NaN in gradients) from “math is right but magnitudes are wrong” (gradient explosion). The brick boundary separates kernel correctness from training dynamics.

Per-transfer. The 3-transfer-per-step contract (C-GPUTRAIN-002) fixes the PCIe budget:

Transfer 1 (H2D): embedding hidden states   ~S×H×4 bytes
Transfer 2 (D2H): logits for cross-entropy  ~S×V×4 bytes
Transfer 3 (H2D): grad_logits to GPU        ~S×V×4 bytes

Any deviation from 3 transfers is a bug, not a tuning knob. For 350M at seq=2048: total ~544 MB/step, overhead ~17 ms on PCIe 4.0 x16 — under 5% of compute time.

12.5.2 Chain of Thought: How Brick Boundaries Diagnose Bugs

When a training run fails, the brick architecture converts “something is broken” into a structured diagnosis:

1. Which granularity? Check per-transfer (D2D size mismatch?), per-block (which layer’s backward fails?), per-kernel (which GEMM overflows?).
2. Local or global? If one block fails and others succeed, the bug is in that block’s kernels. If all blocks succeed but loss diverges, the bug is in training dynamics (LR, grad clipping, optimizer config).
3. Static or dynamic? Buffer overflow is a static invariant violation (detectable by algebraic dimension checking). Gradient explosion is a dynamic stability issue (detectable by runtime sampling).

12.5.3 Five Whys: From Symptom to Root Cause

The brick architecture enforces a disciplined root-cause chain. Concrete example from dogfooding:

This same pattern resolved four bugs during ALB-040 dogfooding:

12.5.4 How Bricks and Contracts Interlock

The gap register (§11) is the feedback loop between profiling and contracts:

Brick profiling finds anomaly
  → File gap (ALB-0XX)
    → Write or update contract obligation
      → Fix upstream brick
        → Verify contract passes (`pv audit`)
          → Dogfood in albor pipeline
            → Close gap

Profiling finds bugs that contracts miss (runtime-only issues like gradient explosion). Contracts prevent bugs that profiling misses (the 50M model’s 2x buffer overflow “worked” through undefined behavior — only a static size invariant would have caught it). Together they form a ratchet: every bug found by profiling becomes a permanent contract obligation that prevents recurrence.

12.6 Verification DAG (Albor End-to-End)

Like the Qwen 3.5 verification DAG in provable-contracts, Albor composes sub-contracts into a full model verification:

softmax ← attention ← gqa
                        ↑
rmsnorm ──────────────── albor-forward ← training-loop
                        ↑                      ↑
gelu ← swiglu ──────────┘                     │
                                               │
rope ──────────────────── albor-forward        │
                                               │
matmul ← gqa                                   │
                                               │
cross-entropy ─────────── training-loss ────────┘
                              ↑
adamw ─────────── optimizer-step ──────── training-loop
                                               │
gradient-accumulation ─────────────────────────┘
                                               │
training-config ─── config-validation ─────────┘
                                               │
knowledge-distillation ── distill-loss ── distill-loop
                              ↑
bpe-tokenizer ─── data-pipeline ─── training-loop

model-merging ─── post-training ─── albor-merged
pruning ────────── post-training ─── albor-pruned

Each node in this DAG is a contract. pv graph contracts/ --format mermaid renders the full dependency graph. A change to any sub-contract triggers re-verification of all dependents.

12.7 Training Stability Contracts

The kernel-level contracts in §12.2 verify local correctness — each kernel produces the right output for its input. They do NOT verify global training stability — that the training loop converges without NaN, that hyperparameters propagate correctly, or that gradients flow to all parameters.

ALB-038, ALB-041, ALB-043, and ALB-044 all passed kernel-level contracts while producing training failures. These contracts bridge the gap between kernel correctness and training correctness.

C-TRAINSTABLE-001: Training Stability

All weights and loss must remain finite for the entire training run.

obligations:
  - "loss.is_finite() for all steps"
  - "weight[i].is_finite() for all i, all steps"
  - "grad[i].is_finite() for all i after clipping, all steps"
falsification: |
  FALSIFY-STABLE-001: Train 100 steps on random init.
  Assert loss.is_finite() at every step.
  Assert no NaN in any model weight after every optimizer step.

C-EMBED-GRAD-001: Activation Gradient Clipping at GPU-CPU Boundary

When GPU backward produces activation gradients that flow to a CPU optimizer, those gradients must be clipped to max_grad_norm before the CPU processes them.

Status: VERIFIED — 350M CUDA test (50 steps) produces zero NaN in embedding weights. Fix in entrenar@86eec38.

motivation: |
  Per-block gradient clipping in CudaGradWorkspace only clips WEIGHT gradients.
  The ACTIVATION gradient in grad_buf_a/b flows unclipped to the CPU embedding
  optimizer. For 24-layer random init, this gradient reaches ~1e35 — overflowing
  the CPU AdamW second moment buffer.
obligation: |
  Before scatter-adding activation gradients into CPU embedding weight gradient:
    grad_norm = L2_norm(activation_grad)
    if grad_norm > max_grad_norm:
        activation_grad *= max_grad_norm / grad_norm
falsification: |
  FALSIFY-EMBEDGRAD-001: Train 350M model (24 layers) for 5 steps.
  Assert embedding weights contain zero NaN values after each optimizer step.

C-HYPERPARAMS-001: Optimizer Hyperparameter Propagation

Every optimizer hyperparameter in the YAML config must reach the actual optimizer constructor. No implicit defaults.

Status: VERIFIED — 350M CUDA test uses explicit AdamW::new() with YAML config values (beta2=0.95, wd=0.1). Fix in entrenar@86eec38.

obligation: |
  For every optimizer in the training loop (GPU AdamW, CPU AdamW, LM head AdamW):
    assert optimizer.lr == config.lr (adjusted for warmup)
    assert optimizer.beta1 == config.beta1
    assert optimizer.beta2 == config.beta2
    assert optimizer.weight_decay == config.weight_decay
    assert optimizer.epsilon == 1e-8 (or config.epsilon if specified)
falsification: |
  FALSIFY-HYPERPARAMS-001: Construct CudaTransformerTrainer with non-default
  YAML config (beta2=0.95, wd=0.1). Verify CPU embed_optimizer.beta2 == 0.95
  and embed_optimizer.weight_decay == 0.1 (not 0.999 and 0.01).
anti_pattern: |
  NEVER: AdamW::default_params(lr)  — hides beta2, wd, epsilon
  ALWAYS: AdamW::new(lr, beta1, beta2, epsilon, wd)  — explicit from config

C-BUFSIZE-001: CUDA Kernel Buffer Size Invariants

Every GPU buffer passed to a CUDA kernel must have algebraically verifiable size matching the kernel’s expected dimensions.

obligation: |
  For gemm_forward(A, B, C, M, K, N):
    assert A.len() >= M * K
    assert B.len() >= K * N
    assert C.len() >= M * N
  For silu_backward(input, grad_output, output):
    assert output.len() >= input.len()
  For rms_norm_backward(input, weight, grad_output, grad_input, grad_weight, S, H):
    assert grad_input.len() >= S * H
    assert grad_weight.len() >= H
falsification: |
  FALSIFY-BUFSIZE-001: Run compute-sanitizer on 10-step 50M training.
  Assert zero illegal address errors.
anti_pattern: |
  NEVER: Reuse a buffer sized for hidden_size as temp for intermediate_size
  ALWAYS: Use dedicated buffers or verify size >= required before kernel call

C-GEMMARGS-001: GEMM Kernel Constructor Argument Ordering

Every GEMM kernel constructor call must pass arguments in the exact order documented by the kernel’s API. Compile-time stride constants baked into PTX are determined by constructor args — wrong order produces wrong strides, not wrong results at the kernel boundary (bounds check passes but data lands in wrong memory).

Status: VERIFIED — 350M CUDA test (50 steps) produces correct backward gradients. Fix in entrenar@846ae0c.

motivation: |
  GemmBackwardAKernel::tiled_unrolled(m, n, k, tile_size) bakes self.n and
  self.k as immediate PTX constants for row/col strides. When called as
  tiled_unrolled(m, k, n, tile) with k and n swapped, the output stride
  becomes vocab_size (32768) instead of hidden_size (512) — writing output
  rows 64× too far apart and overflowing into adjacent GPU allocations.
obligation: |
  For every kernel constructor call:
    assert arg_order matches constructor signature exactly
  Specifically for GEMM backward:
    GemmBackwardAKernel::tiled_unrolled(m, n, k, tile)  # NOT (m, k, n, tile)
    GemmBackwardBKernel::tiled_unrolled(m, n, k, tile)  # NOT (m, k, n, tile)
falsification: |
  FALSIFY-GEMMARGS-001: Train 350M model for 5 steps. Download v_w_k[block0]
  after backward. Assert zero corruption (all values ≥ 0 after optimizer init,
  no values from adjacent buffers).
anti_pattern: |
  NEVER: Guess argument order from variable names (m/n/k are ambiguous)
  ALWAYS: Check constructor signature in trueno-gpu kernel source

C-GPUINIT-001: GPU Buffer Zero Initialization

All optimizer state buffers (m and v for AdamW) must be zero-initialized. GpuBuffer::new() uses cuMemAlloc which returns uninitialized VRAM — the contents are whatever was previously in that memory region.

Status: VERIFIED — All 34 optimizer buffers (18 per-block + 12 LoRA + 4 LM head/norm) zero-initialized via GpuBuffer::from_host(&ctx, &vec![0.0f32; n]). Fix in entrenar@846ae0c.

obligation: |
  For every GpuBuffer used as optimizer state (m, v):
    assert buffer is zero-initialized after allocation
    Use GpuBuffer::from_host(&ctx, &vec![0.0f32; n])
    NOT GpuBuffer::new(&ctx, n)  -- returns uninitialized VRAM
falsification: |
  FALSIFY-GPUINIT-001: Allocate optimizer state, download immediately.
  Assert all values == 0.0.

C-GRADFLOW-001: Gradient Flow Verification

Every trainable parameter must receive a non-zero gradient after one forward+backward step on a non-trivial batch.

obligation: |
  After one forward+backward step on a batch with non-constant inputs:
    for param in model.trainable_parameters():
      assert param.grad().abs().max() > 0
falsification: |
  FALSIFY-GRADFLOW-001: Train 1 step on 50M model with random init.
  Verify all 110 parameter tensors have max(|grad|) > 0.
anti_pattern: |
  NEVER: Create tensors with requires_grad=false in the forward path
  NEVER: Use ops that don't register backward (e.g., manual array copies)
  ALWAYS: Verify gradient flow when adding new layers or ops

C-TRAINCFG-001: Training Configuration Algebraic Consistency

Every training configuration must be algebraically validated BEFORE GPU time is consumed. The epoch/step/data/LR relationship must be provably sufficient.

Status: VERIFIED — ALB-060 config fixed. C-TRAINCFG-001 contract written (contracts/training-config-kernel-v1.yaml), v1 config fixed (epochs: 117), v2 config proven correct (steps_per_epoch = 16994 >= 5000 with expanded 68K dataset). V2 training (ALB-063) reached step ~1183/5000 with loss 10.4→6.9, confirming warmup completes and LR reaches peak 3e-4.

motivation: |
  ALB-060: pretrain-350m.yaml had epochs=1 with 22K sequences and grad_accum=128.
  steps_per_epoch = floor(22079 / 4 / 128) = 43. max_steps=5000 unreachable.
  warmup_steps=2000 never completed. LR peaked at 6.45e-6 (target 3e-4).
  Loss flat at ~10.39 for all 43 steps. Checkpoint contains untrained weights.
  Total wasted: ~12 seconds GPU + debugging time. Contract prevents recurrence.
equations:
  - "steps_per_epoch = floor(num_sequences / batch_size / grad_accum)"
  - "total_achievable_steps = num_epochs × steps_per_epoch"
  - "total_achievable_steps >= max_steps  (HARD REQUIREMENT)"
  - "warmup_steps < total_achievable_steps  (warmup must complete)"
  - "warmup_fraction = warmup_steps / actual_total_steps <= 0.10"
  - "min_epochs = ceil(max_steps / steps_per_epoch)"
  - "total_tokens = actual_steps × batch_size × grad_accum × seq_len"
obligations:
  - "Epoch count sufficient: num_epochs >= ceil(max_steps / steps_per_epoch)"
  - "Warmup completes: warmup_steps < actual_total_steps"
  - "Peak LR reached: exists step t where lr(t) = lr_peak"
  - "Training tokens sufficient: total_tokens >= 10 × num_params"
falsification: |
  FALSIFY-CFG-001: Compute steps_per_epoch for pretrain-350m.yaml.
  With 22079 seqs, batch=4, accum=128: steps_per_epoch=43.
  Assert 1 × 43 < 5000 (proves epochs=1 is insufficient).
  FALSIFY-CFG-002: Assert warmup_steps (2000) > total_steps (43)
  (proves warmup never completes with epochs=1).

Full contract: contracts/training-config-kernel-v1.yaml — 7 equations, 8 proof obligations, 5 falsification tests, 2 Kani harnesses.

C-STREAMSYNC-001: Stream Synchronization Before D2H Transfers

Every cuMemcpyDtoH (or copy_to_host_at()) call that reads data written by GPU kernels on a non-default stream MUST be preceded by stream.synchronize().

motivation: |
  ALB-065: gradient clipping downloaded 9 GPU buffers via cuMemcpyDtoH
  without stream synchronization. trueno CudaStream uses CU_STREAM_NON_BLOCKING;
  cuMemcpyDtoH only synchronizes with the default stream. Backward kernels
  hadn't finished → garbage clip scale → NaN → silent SIGABRT (process death
  with no error output). Training was stable with CUDA_LAUNCH_BLOCKING=1 but
  crashed within 15 seconds without it.
obligation: |
  stream.synchronize() MUST precede every cuMemcpyDtoH that reads kernel output.
  No exceptions. The sync ensures all prior kernel launches have completed.
falsification: |
  FALSIFY-GPU-008: Run 350M training for 50+ steps WITHOUT CUDA_LAUNCH_BLOCKING=1.
  Verify process stays alive, loss is finite, no CUDA errors in dmesg/Xid log.
anti_pattern: |
  NEVER: call copy_to_host_at() after kernel launches without stream.synchronize()
  NEVER: rely on cuMemcpyDtoH to synchronize non-blocking streams (it doesn't)
  DIAGNOSTIC: if training crashes without CUDA_LAUNCH_BLOCKING=1 but works with it,
  this is the FIRST contract to check

Full contract: contracts/training-gpu-kernel-v1.yaml — stream_synchronization equation + proof obligation.

12.7.1 Observability Discipline

All training observability MUST use the renacer tracing infrastructure.

entrenar integrates renacer in src/run.rs (span lifecycle: create_span, emit_metric_event, end_span). The src/monitor/drift.rs module provides anomaly detection (DriftStatus, AnomalySeverity) that can automatically flag NaN, gradient explosion, and loss divergence.

obligation: |
  NEVER: eprintln!(), println!(), dbg!() for training diagnostics
  ALWAYS: tracing::debug!(), tracing::warn!() with structured fields
  ALWAYS: emit_metric_event() for training metrics (loss, grad_norm, lr)
motivation: |
  Ad-hoc eprintln! creates cleanup debt, is invisible to tracing infra,
  loses brick profiling boundary isolation, and cannot be filtered at runtime.
  renacer BrickTracer provides structured, filterable, permanent observability.

13. pmat Compliance & Quality Gates

13.1 Scope: Where Quality Applies

Albor is a project repo (configs, scripts, contracts, docs). It produces no Rust library code. All quality gates apply to upstream Rust changes made in service of Albor’s gaps — not to albor’s shell scripts or YAML configs.

# Run on all modified stack components (NOT on albor itself)
pmat comply check --strict ../aprender      # ALB-001, 006, 009, 011
pmat comply check --strict ../entrenar      # ALB-003, 004
pmat comply check --strict ../trueno        # ALB-005
pmat comply check --strict ../realizar      # ALB-010
pmat comply check --strict ../alimentar     # ALB-007, 018, 019, 020
pmat comply check --strict ../repartir      # ALB-002, 008

13.2 Quality Gate Thresholds (Upstream Rust Code)

13.3 Quality Gate Thresholds (Albor Repo)

13.3 pmat Quality Commands for Albor

# TDG analysis of all Albor-touched code
pmat analyze tdg ../aprender --include-components
pmat analyze tdg ../entrenar --include-components

# Find coverage gaps (highest ROI targets)
pmat query --coverage-gaps --limit 30 --exclude-tests

# Fault pattern audit (unwrap, panic, unsafe)
pmat query "training" --faults --exclude-tests

# Full quality audit on distillation code
pmat query "distill" --churn --duplicates --entropy --faults -G

# Complexity check on new kernels
pmat query "knowledge_distillation" --max-complexity 15 --include-source

# Create quality baseline before Albor work begins
pmat tdg baseline create

# Check for regressions after each phase
pmat tdg check-regression --baseline

13.5 Certeza Three-Tier Testing (Upstream Repos)

When modifying upstream Rust code for gap fixes, follow certeza tiers:

Tier 1: On-Save (sub-second)

cargo check && cargo test --lib -- --quiet    # Type check + unit tests

Tier 2: On-Commit (1-5 minutes)

cargo test                                     # Full test suite
cargo llvm-cov --summary-only                  # Coverage ≥ 95%
pmat analyze tdg                               # TDG regression check
pv audit contracts/ --binding                  # Contract compliance

Tier 3: On-Merge / Nightly (hours)

cargo mutants --no-times                       # Mutation score ≥ 85%
cargo kani                                     # Formal verification
batuta falsify . --min-grade toyota-standard   # 108-item checklist
pmat rust-project-score --full                 # Comprehensive quality score

13.6 Albor Pipeline Commands

Since albor is a project repo, its primary interface is apr pipeline. No Makefiles, no shell scripts. One manifest, one DAG.

# ── Pipeline (the only entry point you need) ──
apr pipeline plan configs/pipeline/albor.yaml     # Full DAG dry-run (no GPU, no writes)
apr pipeline apply configs/pipeline/albor.yaml    # Execute everything (resumable)
apr pipeline status                               # What's converged / pending / failed
apr pipeline drift                                # Detect unauthorized state changes

# ── Targeted execution (run one step + its dependencies) ──
apr pipeline apply configs/pipeline/albor.yaml --target train-350m
apr pipeline apply configs/pipeline/albor.yaml --target eval-code
apr pipeline apply configs/pipeline/albor.yaml --target publish

# ── Force re-run (ignore converged state) ──
apr pipeline apply configs/pipeline/albor.yaml --target distill --force

# ── Individual subcommands (for development / debugging) ──
apr train plan configs/train/pretrain-350m.yaml   # Plan one step standalone
apr train apply configs/train/pretrain-350m.yaml --seed 42
apr monitor ./checkpoints/albor-base-350m/        # Live TUI
apr experiment view --db .entrenar/experiments.db  # Browse experiments

# ── Quality (upstream repos — run independently of pipeline) ──
pmat tdg baseline create                          # TDG baseline across all components
pmat comply check --strict ../aprender
pmat comply check --strict ../entrenar
pv validate contracts/*.yaml                      # Contract schema validation
pv status contracts/                              # Contract completeness
batuta falsify . --min-grade toyota-standard      # 108-item falsification checklist
# Current score: 100.0% (108/108 PASS) — achieved 2026-03-04

14. Batuta Falsification Checklist

14.1 108-Item Popperian Assessment

The Albor project itself is subject to batuta’s 108-item falsification checklist:

# Full assessment
batuta falsify . --verbose --format markdown --output docs/falsification-report.md

# Critical-only (blocks release)
batuta falsify . --critical-only

# CI-friendly output
batuta falsify . --format github-actions --min-grade kaizen-required

14.2 Key Sections Applied to Albor

Section 1: Sovereign Data Governance (SDG)

• All training data has documented provenance (HuggingFace commit SHAs)
• No PII in training corpus (alimentar quality check)
• Data residency: all data stored on owned hardware (lambda + intel)
• Teacher model license verified (Apache 2.0)

Section 3: Hypothesis-Driven Development (HDD)

• Each improvement stage has a falsifiable hypothesis:
  • “Distillation improves avg benchmark by >5%” (FALSIFY-ALBOR-005)
  • “Pruning at 50% sparsity degrades benchmarks by <2%” (FALSIFY-ALBOR-008)
  • “Q4 quantization degrades perplexity by <5%” (FALSIFY-ALBOR-009)
• Reproducibility standard: Gold (deterministic seeds, versioned data, BLAKE3 checkpoint hashes, Cargo.lock pinning)

Section 4: Numerical Reproducibility (NR)

• Float determinism enforced via fixed seeds and operator ordering
• Cross-platform consistency: checkpoint trained on lambda loads on intel
• SIMD parity: all kernels have provable-contracts SIMD equivalence obligations

Section 5: Performance & Waste Elimination (PW)

• Seven Wastes (Muda) applied to training pipeline:
  • No redundant data copies (alimentar streaming)
  • No idle GPU time (pre-computed teacher logits)
  • No over-processing (progressive model sizing: 50M → 125M → 350M)

Section 6: Safety & Formal Verification (SF)

• Critical kernels have Kani proofs (softmax, attention, cross-entropy)
• New kernels (KD loss, gradient accumulation) get Kani harnesses

Section 10: Architectural Invariants (AI) — CRITICAL

• AI-01: All model operations use apr (no manual weight manipulation)
• AI-02: Every checkpoint is BLAKE3-hashed and version-tracked
• AI-03: Training config is immutable once committed (no runtime overrides)
• AI-04: Eval results are reproducible (fixed seed, deterministic batching)
• AI-05: No undeclared dependencies (Cargo.lock enforced)

14.3 Current Grade

Perfect Score: 100.0% (108/108 PASS) — achieved 2026-03-04.

This exceeds the Toyota Standard (90-100%) target:

• All 5 Critical items pass (Section 10)
• All Major items pass
• All Minor items pass
• Zero PARTIAL, zero FAIL

Score progression across 14 MLOps survey batches: 34% → 100% (see entrenar/docs/specifications/world-class-mlops-survey.md).

15. Implementation Phases

Phase 0: Pipeline Manifest, Contracts & Quality Baseline (Week 1)

• Write configs/pipeline/albor.yaml — full pipeline manifest (infra + data + train + eval + publish)
• apr pipeline plan — validate entire DAG, estimate resources
• apr pipeline apply --target cuda-driver --target vulkan-driver --target data-dir — provision infra
• Verify trueno wgpu on W5700X via Vulkan (not Metal — Linux)
• Verify trueno CUDA on 4090
• Download Qwen3-Coder-Next to intel box, verify it loads in realizar
• pmat tdg baseline create on all stack components
• pv coverage contracts/ --binding — establish contract coverage baseline
• batuta falsify . --critical-only — initial falsification assessment

Phase 1: Data Pipeline + Tokenizer Contract (Week 1-2)

• Ingest local ground truth corpora via alimentar import local (fix ALB-019 if needed)
  • depyler: examples/ + tdd-book/tests/ (~1,845 files, ~219K lines)
  • hf-ground-truth-corpus (~11,928 files)
  • jax-ground-truth-corpus (~2,697 files)
  • vllm-ground-truth-corpus (~1,118 files)
• Ingest local ML framework code (Tier 2, ~53K files)
• Download external datasets via alimentar import hf (StarCoder Python, FineWeb-Edu)
• Quality validation via alimentar quality check on all sources
• Build weighted training mix with 10x upsampling on Tier 1 (fix ALB-020 if needed)
• Write bpe-tokenizer-kernel-v1.yaml contract (ALB-014)
• pv probar + pv kani on tokenizer contract
• Train BPE tokenizer on mixed corpus (fix ALB-001 if needed)
• Verify FALSIFY roundtrip: decode(encode(text)) = text for all test data
• Tokenize all data into sharded Parquet
• Apply FIM transforms to code sequences (fix ALB-018 if needed)
• Create train/val/test splits via alimentar
• Record SHA-256 hashes + provenance manifest for all data artifacts
• pmat comply check --strict on alimentar changes

Phase 2: Pipeline Validation — 50M Model (Week 2) – COMPLETE

• Write gradient-accumulation-kernel-v1.yaml contract (ALB-017)
• Write configs/train/pretrain-50m.yaml (model arch + training + monitoring)
• Train albor-50M on 4090 — 500 rows, 31 steps, 110.7s, loss 10.3→4.42
• Validate apr monitor — ALB-025 FIXED (presentar widget migration complete)
• Validate Andon alerts during full training run
• Fix ALB-009 FIXED
• Verify FALSIFY-ALBOR-001 (loss decreases) — CORROBORATED
• Verify FALSIFY-ALBOR-002 (gradient bounds) — per-step logging now available (ALB-035 FIXED)
• pv audit — PASS: 7/7 contracts, 0 findings
• Milestone: Training loop converges ✓, contracts pass ✓

Phase 3: Base Model — 350M Pre-Training (Week 2-4) – IN PROGRESS

• Write configs/train/pretrain-350m.yaml — pre-tokenized ByteLevel BPE v2, 22K×2048 tokens
• Train albor-base-350m on 4090 — STARTED (2760 batches, ~20h est.)
• Build evaluation infrastructure — eval-code.py, eval-perplexity.py, 35 benchmark problems
• Fix ALB-038 FIXED — RMSNorm + attention backward ops, all 20 params receive gradients
• Fix ALB-041 FIXED — D2D buffer size mismatch in backward_attention (entrenar@a48e3d2)
• Fix ALB-043 FIXED — backward_ffn buffer overflow + SwiGLU gradients (entrenar@f7805f1)
• Fix ALB-044 FIXED — activation gradient clipping at GPU-CPU boundary + CPU optimizer hyperparams (entrenar@86eec38)
• Fix ALB-059 FIXED — GEMM backward constructor args n/k swapped, buffer overflow into optimizer states + zero-init optimizer m/v (entrenar@846ae0c)
• Write training-memory-kernel-v1.yaml contract (ALB-039) — VRAM budget estimation
• Write training-gpu-kernel-v1.yaml contract (ALB-040) — GPU-resident training invariants
• Implement CudaTransformerTrainer (ALB-040) — 3 PCIe transfers/step vs ~16K
• Dogfood CUDA training — 50M test: 3 steps, loss 10.4→11.7, GPU forward+backward working
• ALB-037 FIXED — realizar loads trained SafeTensors checkpoint, generates tokens (e2e verified)
• 350M CUDA test training — 50 steps, loss 10.39→5.92 (best 5.53), checkpoint valid
• realizar inference verified — 218 tensors loaded, generates from trained weights
• Checkpoint validation: PASS (weights trained, not initialization)
• Perplexity eval: 31,926 (finite, consistent with 50-step model — random baseline ~32,768)
• Fix ALB-060 CONFIG FIXED — epochs=1 only ran 43/5000 steps. C-TRAINCFG-001 contract written. Config fixed (v1: epochs=117, v2: epochs=1 with 68K seqs)
• Expand training data: Tier 1 10x + 8 Tier 2 repos → v2 dataset (67,977 seqs, 139M tokens)
• Fix ALB-071 FIXED — embed gradient clipping decoupled from weight grad_clip (entrenar@d07d67d)
• Fix ALB-072 FIXED — fp16 loss scaling (65536x) removed from fused CE kernel; all backward uses f32, no underflow risk (entrenar@44d3e74)
• Full 350M v2 training — reached step 1183/5000, loss 10.40→6.85, val_ppl=1008. Crashed: ALB-073 (PTX selp) + ALB-074 (buffer overflow from stale binary). Step 1000 checkpoint saved (1520 MB).
• Fix ALB-073 FIXED — fused_cross_entropy selp arg order, same class as ALB-069 (trueno@10bec89)
• Fix ALB-074 FIXED — stale binary missed eval truncation fix. Rebuilt with entrenar@5c4c2d8.
• Monitor training via apr monitor (ALB-025 FIXED)
• Data scaling: Download codeparrot-clean (2M files, ~4.4B tokens) → pretokenize at 1024 → ~5.2M sequences
• Full 350M v3 training — PENDING: 250K steps on ~1B tokens from codeparrot-clean. Config: pretrain-350m-v3.yaml. ETA ~10 days.
• Validate loss curve, perplexity convergence
• HumanEval pass@1 evaluation (target >8%)
• Verify FALSIFY-ALBOR-003 (checkpoint determinism)
• pmat tdg check-regression on all touched components
• Milestone: HumanEval pass@1 > 8%, Perplexity < 30, TDG grade A maintained

Phase 4: Teacher Setup & Logit Pre-Computation (Week 3-5)

• Fix ALB-010: Add Qwen3-Coder-Next support to realizar (stretch — 3-4 week blocker)
• Download Qwen2.5-Coder-3B interim teacher (5.75 GiB, Apache 2.0) — unblocks distillation without ALB-010
• Validate 3B teacher: apr distill --stage precompute works, RosettaStone handles sharded SafeTensors
• Create distillation config: configs/train/distill-qwen3b.yaml (T=4.0, α=0.5, LoRA r=16)
• Validate teacher inference on intel (CPU, fp16, 300GB RAM) — for 80B stretch goal
• Write knowledge-distillation-kernel-v1.yaml contract (ALB-013) — DOGFOODING
• pv kani on KD loss contract (KL non-negativity, temperature scaling)
• Fix ALB-011 FIXED — apr distill --config --stage precompute|train works
• Pre-compute 3B teacher logits on v2 dataset (background, 4-8h CPU)
• Verify FALSIFY-ALBOR-006 (teacher logit integrity)
• Store as sharded Parquet via alimentar
• pmat comply check --strict on realizar changes
• Milestone: Teacher logits verified, KD contract at Level 4

Phase 5: Knowledge Distillation (Week 5-6)

• Implement apr distill apply with KD loss
• Distill albor-base-350m → albor-distill-350m
• Verify FALSIFY-ALBOR-004 (KL non-negativity in production)
• Verify FALSIFY-ALBOR-005 (distillation improves benchmarks)
• Benchmark: measure improvement over base
• pv probar --binding on KD contract with actual training data
• Milestone: >5% avg benchmark improvement, KD contract fully wired

Phase 6: Post-Training Optimization (Week 6-8)

• Write model-merging-kernel-v1.yaml contract (ALB-015) — DOGFOODING
• Write pruning-kernel-v1.yaml contract (ALB-016) — DOGFOODING
• Fine-tune with LoRA: apr finetune → albor-instruct
• Merge variants: apr merge --method slerp → albor-merged
• Verify FALSIFY-ALBOR-007 (SLERP interpolation bound)
• Prune: apr prune --method wanda → albor-pruned
• Verify FALSIFY-ALBOR-008 (sparsity guarantee)
• Quantize: apr quantize --method q4_k → albor-q4
• Verify FALSIFY-ALBOR-009 (quantization fidelity)
• Benchmark every variant
• pv coverage contracts/ --binding — final contract coverage report
• Milestone: Full ladder complete, all post-training contracts pass

Phase 7: Quality Assurance & Falsification Sweep (Week 8)

• batuta falsify . --min-grade toyota-standard --verbose — full 108-item assessment
• pmat rust-project-score --full on all touched components
• pmat tdg check-regression --baseline — no quality regressions
• pv graph contracts/ --format mermaid — publish verification DAG
• pv status contracts/ — all contracts at Level 3+, critical at Level 4
• cargo mutants --no-times on all new code — mutation score ≥ 85%
• cargo llvm-cov — coverage ≥ 95% on all new code
• Address any falsification failures or contract violations
• Milestone: Toyota Standard grade, all quality gates green

Phase 8: Evaluation, Leaderboard Submission & Publication (Week 8-9)

• Final eval on all benchmark tasks (all 6 model variants)
• Run bigcode-evaluation-harness with leaderboard-standard params on best model
• Submit PR to Big Code Models Leaderboard (community_results/ folder)
• Export all models: SafeTensors + GGUF
• apr publish to HuggingFace Hub as paiml/albor-*
• Write model card with full reproducibility details + leaderboard results
• Publish training logs, loss curves, eval trajectories
• Publish verification report (contract status, falsification results)
• batuta falsify . --format markdown --output docs/falsification-report.md
• Milestone: Models on HuggingFace, leaderboard submission live, quality evidence published

Phase 9: Distributed Training — Stretch (Week 9+)

• entrenar native DDP infrastructure (TCP wire protocol v2, GradientServer, WorkerClient, PerBlockGradientAccumulator, RingAllReduce) — entrenar#133
• Wire DDP train_batch() into DistributedCudaTrainer — COMPLETE (train_loop_cuda_distributed, allreduce_impl, spawn_coordinator_thread)
• Multi-process launcher — COMPLETE (rank 0 auto-spawns GradientServer, all ranks connect as WorkerClient via --distributed CLI flags)
• wgpu backward pass in trueno (ALB-005) — for cross-vendor GPU support
• Full distributed training: 4090 + W5700X x2
• Milestone: Multi-GPU training demonstrated

16. Reproducibility Protocol

Every artifact in the albor pipeline is reproducible from source. This chapter documents the exact commands, seeds, and checksums needed to reproduce the full training pipeline from raw code corpora to trained model.

16.1 Artifact Tracking

16.2 Full Reproduction Commands

Step 1: Corpus Preparation

v1 pipeline (Tier 1 only, 17K rows):

# Import Tier 1 ground truth corpora
alimentar import local /path/to/depyler -o data/raw/depyler.parquet
alimentar import local /path/to/hf-ground-truth-corpus -o data/raw/hf.parquet
alimentar import local /path/to/jax-ground-truth-corpus -o data/raw/jax.parquet
alimentar import local /path/to/vllm-ground-truth-corpus -o data/raw/vllm.parquet

# Mix training split (weighted sampling)
alimentar mix \
    data/raw/depyler.parquet:0.4 \
    data/raw/hf.parquet:0.3 \
    data/raw/jax.parquet:0.15 \
    data/raw/vllm.parquet:0.15 \
    -o data/tokenized/train/mixed.parquet \
    --seed 42

v2 pipeline (Tier 1 10x + 8 Tier 2 repos, 45K rows → 68K sequences):

# Convert Tier 2 source repos to Parquet (alimentar can't read source dirs)
for repo in pytorch hf-repos mlflow vllm-full tgi algo-corpus cuda-python llms-with-hf; do
    python3 scripts/source-to-parquet.py ~/src/$repo $repo data/parquet/tier2/$repo.parquet
done

# Mix Tier 1 (10x upsampled) + Tier 2 (1x)
alimentar mix \
    data/parquet/depyler/shard_0000.parquet:10.0 \
    data/parquet/hf-ground-truth/shard_0000.parquet:10.0 \
    data/parquet/jax/shard_0000.parquet:10.0 \
    data/parquet/vllm/shard_0000.parquet:10.0 \
    data/parquet/tier2/pytorch.parquet:1.0 \
    data/parquet/tier2/hf-repos.parquet:1.0 \
    data/parquet/tier2/mlflow.parquet:1.0 \
    data/parquet/tier2/vllm-full.parquet:1.0 \
    data/parquet/tier2/tgi.parquet:1.0 \
    data/parquet/tier2/algo-corpus.parquet:1.0 \
    data/parquet/tier2/cuda-python.parquet:1.0 \
    data/parquet/tier2/llms-with-hf.parquet:1.0 \
    -o data/staging/mixed-expanded.parquet --seed 42

# Apply FIM (50% PSM)
alimentar fim data/staging/mixed-expanded.parquet \
    -o data/staging/mixed-expanded-fim.parquet --rate 0.5 --format psm --seed 42

Step 2: Tokenizer Training

# v1 tokenizer (whitespace-split BPE — has ALB-036 limitation)
apr tokenize apply \
    --data data/staging/corpus-raw.txt \
    --vocab-size 32768 \
    --algorithm bpe \
    -o models/albor-tokenizer/ \
    --max-lines 100000

# v2 tokenizer (ByteLevel BPE — preserves whitespace)
python scripts/train-tokenizer-v2.py \
    --corpus data/staging/corpus-raw.txt \
    --vocab-size 32768 \
    --output models/albor-tokenizer-v2/

Step 3: Pre-Tokenization

# Pre-tokenize full training data (v2 tokenizer, 2048-token chunks)
python scripts/pretokenize.py \
    --input data/tokenized/train/mixed.parquet \
    --tokenizer models/albor-tokenizer-v2/tokenizer.json \
    --seq-len 2048 \
    --output data/pretokenized-2048/train/train.parquet

# Pre-tokenize validation data
python scripts/pretokenize.py \
    --input data/tokenized/val/val.parquet \
    --tokenizer models/albor-tokenizer-v2/tokenizer.json \
    --seq-len 2048 \
    --output data/pretokenized-2048/val/val.parquet

Step 4: Model Training

# 50M pipeline validation (< 2 minutes)
make train-50m
# Equivalent to:
# apr train apply --task pretrain --config configs/train/pretrain-50m.yaml

# 350M base model, v2 data (~20 hours on RTX 4090)
apr train apply --task pretrain --config configs/train/pretrain-350m-v2.yaml
# v2 config: epochs=38, warmup=500, 67977 seqs, 5000 max_steps
# C-TRAINCFG-001 verified: steps_per_epoch=132, 38×132=5016 >= 5000

# Legacy v1 (22K seqs, fixed epochs=117 post ALB-060)
# apr train apply --task pretrain --config configs/train/pretrain-350m.yaml

Step 5: Checkpoint Conversion (for evaluation)

# Convert entrenar 1D-flat SafeTensors to realizar 2D format
python scripts/convert-checkpoint.py checkpoints/albor-base-350m/ \
    --config configs/train/pretrain-350m.yaml

Step 6: Evaluation

# Validate all benchmarks (no model needed)
make eval-validate

# Perplexity evaluation (needs trained model)
make eval-perplexity-350m

# Monitor active training
make training-status

16.3 Key SHA-256 Checksums

See docs/PROVENANCE.md for complete checksums. Key artifacts:

16.4 Verification

# Verify data checksums
sha256sum data/tokenized/train/mixed.parquet
sha256sum data/pretokenized-2048/train/train.parquet
sha256sum models/albor-tokenizer-v2/tokenizer.json

# Verify training config reproducibility
apr train plan --task pretrain --config configs/train/pretrain-350m.yaml

# Verify contract integrity
pv validate contracts/*.yaml
pv coverage contracts
pv audit contracts/*.yaml

17. Success Criteria

Minimum Viable (Phase 3 complete)

• 350M base model trained on 4090 to convergence (target: ~10B tokens; current: 139M v2 dataset)
• FIM (fill-in-the-middle) training implemented and validated (ALB-018 FIXED — alimentar fim verified)
• HumanEval pass@1 > 8% (baseline Python capability, beat random)
• HumanEval-FIM working (model can infill Python code)
• Entire pipeline uses only sovereign stack components
• All training artifacts reproducible from spec
• All existing kernel contracts pass pv audit (Level 2+)
• pmat comply check passes on all modified components

Current blockers for Phase 3 completion:

• ALB-038 (Critical): entrenar saves initialization weights, not trained weights FIXED (entrenar@91ba9da, @1ede409)
• ALB-035: No per-step loss logging during training FIXED (entrenar@5d41a96)
• ALB-041: D2D buffer mismatch in backward_attention FIXED (entrenar@a48e3d2)
• ALB-037: realizar ignores loaded weights FIXED (e2e verified: realizar run loads 350M trained checkpoint, generates tokens from 218 tensors)
• ALB-043 (Critical): backward_ffn buffer overflow + missing SwiGLU gradients FIXED (entrenar@f7805f1)
• ALB-044 (Critical): activation gradient clipping + CPU optimizer hyperparams FIXED (entrenar@86eec38)
• ALB-059 (Critical): GEMM backward constructor n/k swapped — buffer overflow into optimizer states FIXED (entrenar@846ae0c)
• ALB-040: GPU-resident pretraining VERIFIED — 350M CUDA test: 50 steps, loss 10.39→5.92, checkpoint valid, realizar inference works
• ALB-042: CUDA runtime errors produce silent loss=0.0 — OPEN (workaround: CUDA_VISIBLE_DEVICES="")
• ALB-069 (Critical): PTX selp_f32 argument order in fused cross-entropy FIXED (trueno@10bec89)
• ALB-060 (Critical): Training ran only 43/5000 steps (epochs=1). CONFIG FIXED: C-TRAINCFG-001 contract + v2 config. V2 training (ALB-063) restarted after ALB-069 fix — PID 106929, loss=10.39 at step 1.

350M CUDA test results (50 steps, post ALB-059 fix):

• Loss: 10.39 → 5.92 (best: 5.53) — clear convergence with correct GEMM backward
• Training time: ~400s (~8s/step) with PTX; ~26s (~0.5s/step) with cuBLAS (ALB-075/077)
• Checkpoint: 1.59 GB SafeTensors, 218 tensors, config.json saved
• Checkpoint validation: PASS (weights trained, layers distinct)
• realizar inference: loads model, generates tokens (gibberish at 50 steps — expected)
• Perplexity: 31,926 (finite; random baseline ~32,768 for vocab 32K)

350M v3 training (250K steps, codeparrot-clean, ALB-077 fix) — STOPPED:

• Final: step 28K, loss=6.43, val_ppl=1018, 6.7K tok/s, 19.3% MFU
• Plateau since step 12K — val_ppl stalled at ~1000, gnorm collapsed 3.0→0.13
• Root cause: ALB-079 (constant lr after warmup, no cosine decay) + ALB-080 (4K tokens/step, 48-128x too small)
• Checkpoints: step 1K-28K (1520 MB each, all verified OK)
• No NaN in 28K steps (ALB-077: tensor cores disabled, CUBLAS_DEFAULT_MATH)

350M v4 training (ALB-079 + ALB-080 fixes) — RESUMED from step 500:

• Fixes: cosine LR decay (entrenar PR #241) + gradient_accumulation=32 (131K tokens/step)
• Original run: 500 steps, val_ppl=1032.7 (matched v3 at 57% token budget)
• System reboot at step 553; resumed from step-500 checkpoint
• Extended resume: step 350 (cum. step 850), best loss=5.69 at step 262
• 111M tokens processed (2.1% of 5.3B available); loss plateau at mean ~6.65
• Cosine decay just engaging (lr 3.00e-4→2.98e-4); expect plateau break at step 1000+
• ZClip catching gradient spikes (z=2.0–4.0), gnorm healthy 0.05–0.32
• Throughput: 3,564–3,569 tok/s steady, 10.3% MFU, 14-16 GB / 24 GB VRAM
• Target: val_ppl < 100 by 1B tokens (~60 hours remaining)
• Same hardware (RTX 4090), same data (codeparrot-clean, 5.3B tokens available)

Good (Phase 5 complete)

• Distillation from Qwen3.5-35B-A3B demonstrated (ALB-010); fallback: Qwen2.5-Coder-3B (dense)
• albor-distill-350m outperforms albor-base-350m on all code benchmarks
• HumanEval pass@1 > 15% (beat CodeGen-350M-mono’s 12.8% via distillation from 35B MoE teacher)
• MBPP pass@1 > 12%
• FIM infill working (qualitatively: model can complete Python between prefix and suffix)
• KD contract at Level 4 (Kani-proved KL non-negativity)
• All FALSIFY-ALBOR tests pass (001-006)

Full Success (Phase 8 complete)

• All 6 model variants benchmarked (base → distill → instruct → merged → pruned → q4)
• Benchmark trajectory published showing improvement at each stage
• Submitted to Big Code Models Leaderboard — first sub-1B model on the board
• Q4 model: <50ms/token on CPU, <10ms/token on GPU (code completion latency)
• Critical path gaps (ALB-001, 006, 009, 011, 018) closed with upstream fixes; ALB-010 (Qwen3.5-35B-A3B MoE inference) PR #133 MERGED, weight loading remaining
• Models published on HuggingFace as paiml/albor-python-*
• Q4 quantized model < 100MB, runs on consumer hardware
• All 8 kernel contracts written and verified (ALB-013–017, ALB-039–040, ALB-060)
• batuta falsify: Toyota Standard grade (≥90/108) — ACHIEVED: 100% (108/108 PASS)
• pmat TDG: Grade A on all touched components
• Test coverage ≥ 95%, mutation score ≥ 85% on all new code
• All 9 FALSIFY-ALBOR tests pass
• Verification DAG published via pv graph

Stretch Goals

• HumanEval pass@1 > 20% (strong distillation result at 350M)
• DS-1000 pass@1 > 10% (data science code generation)
• Editor integration: VS Code / Neovim / Helix extension using realizar as backend
• Distributed gradient-parallel training across 4090 + W5700X demonstrated (entrenar DDP #133 infra in place)
• apr pipeline apply reproduces entire ladder from bare metal to published model
• BabyLM 2026 submission using constrained data variant
• All critical kernels at Level 4 (Kani formal proofs)
• Lean 4 theorem stubs generated for core training loop invariants

18. Reference Commands

# ═══════════════════════════════════════════════════════════
# THE PIPELINE (two orchestrators working together)
# ═══════════════════════════════════════════════════════════

# Infrastructure provisioning (forjar — bare metal to ready state)
forjar validate -f configs/pipeline/infra-only.yaml   # Validate
forjar apply -f configs/pipeline/infra-only.yaml       # Provision

# ML pipeline orchestration (batuta playbook — data to published model)
batuta playbook validate configs/pipeline/albor-playbook.yaml  # Validate DAG
batuta playbook run configs/pipeline/albor-playbook.yaml       # Execute (resumable)
batuta playbook status configs/pipeline/albor-playbook.yaml    # Check progress

# Unified pipeline (apr pipeline wraps forjar + batuta)
apr pipeline plan configs/pipeline/albor.yaml
apr pipeline apply configs/pipeline/albor.yaml
apr pipeline status

# ═══════════════════════════════════════════════════════════
# DATA PIPELINE
# ═══════════════════════════════════════════════════════════

# Import local codebases
alimentar import local /path/to/codebase -o data/raw/corpus.parquet

# Weighted mix with upsampling
alimentar mix a.parquet:0.4 b.parquet:0.3 c.parquet:0.15 d.parquet:0.15 \
    -o data/tokenized/train/mixed.parquet --seed 42

# FIM transform
alimentar fim data.parquet -o data-fim.parquet --rate 0.5 --format psm

# Quality profiles
alimentar quality profiles

# ═══════════════════════════════════════════════════════════
# TOKENIZER
# ═══════════════════════════════════════════════════════════

# v1: BPE with apr (whitespace-split — ALB-036 limitation)
apr tokenize plan --data corpus.txt --vocab-size 32768
apr tokenize apply --data corpus.txt --vocab-size 32768 --algorithm bpe -o tokenizer/

# v2: ByteLevel BPE with Python (recommended — preserves whitespace)
python scripts/train-tokenizer-v2.py --corpus corpus.txt --vocab-size 32768 \
    --output models/albor-tokenizer-v2/

# Pre-tokenize for training (bypasses tokenizer format gap ALB-033)
python scripts/pretokenize.py --input data.parquet \
    --tokenizer models/albor-tokenizer-v2/tokenizer.json \
    --seq-len 2048 --output data/pretokenized-2048/train/train.parquet

# ═══════════════════════════════════════════════════════════
# TRAINING
# ═══════════════════════════════════════════════════════════

# Plan (dry-run, validate config)
apr train plan --task pretrain --config configs/train/pretrain-350m.yaml

# Train (execute)
apr train apply --task pretrain --config configs/train/pretrain-350m.yaml

# Makefile shortcuts
make train-50m        # ~2 min on RTX 4090
make train-350m       # ~20 hours on RTX 4090
make training-status  # Check running training

# ═══════════════════════════════════════════════════════════
# EVALUATION
# ═══════════════════════════════════════════════════════════

# apr eval (perplexity — ALB-037 FIXED, realizar loads checkpoints)
apr eval checkpoints/albor-base-350m/model.safetensors \
    --dataset custom --text "def foo():" --threshold 30

# Python eval scripts (supplement)
python scripts/eval-code.py configs/eval/humaneval-subset.jsonl --validate-only
python scripts/eval-code.py configs/eval/humaneval-subset.jsonl --api http://localhost:8080
python scripts/eval-perplexity.py checkpoints/albor-base-350m/ \
    --data data/pretokenized-2048/val/val.parquet --seq-len 2048 --threshold 30

# Convert entrenar checkpoint for realizar
python scripts/convert-checkpoint.py checkpoints/albor-base-350m/ \
    --config configs/train/pretrain-350m.yaml

# Makefile shortcuts
make eval-validate           # Validate all benchmark canonical solutions
make eval-perplexity-350m    # Run perplexity eval

# ═══════════════════════════════════════════════════════════
# MONITORING (run in a separate terminal during training)
# ═══════════════════════════════════════════════════════════

bash scripts/monitor-training.sh                     # Training process + GPU + log
apr monitor ./checkpoints/albor-base-350m/           # Live training TUI (ALB-025 FIXED)
apr experiment view --db .entrenar/experiments.db     # Browse past experiments

# ═══════════════════════════════════════════════════════════
# POST-TRAINING (Phases 4-6)
# ═══════════════════════════════════════════════════════════

# Distillation
apr distill --config configs/train/distill.yaml --plan
apr distill --config configs/train/distill.yaml --stage precompute
apr distill --config configs/train/distill.yaml --stage train

# Fine-tuning
apr finetune --plan --model-size 350M --vram 24 --method lora --rank 16

# Model operations
apr merge a.safetensors b.safetensors --strategy slerp -o merged.safetensors
apr prune model.safetensors --method wanda --sparsity 0.5 -o pruned.safetensors
apr quantize model.safetensors --method q4_k -o model.gguf
apr export model.safetensors --format gguf -o model.gguf
apr publish checkpoints/albor-350m/ paiml/albor-base-350m

# ═══════════════════════════════════════════════════════════
# QUALITY (bashrs is KING of linting)
# ═══════════════════════════════════════════════════════════

# bashrs — sovereign linter for all shell artifacts
bashrs make lint Makefile                          # Makefile quality
bashrs classify Makefile                           # Safety classification
bashrs make purify Makefile                        # Deterministic output

# provable-contracts — kernel correctness
pv validate contracts/*.yaml                       # Contract schemas
pv coverage contracts                              # Obligation coverage
pv generate contracts/*.yaml                       # Scaffold + tests + harnesses
pv book contracts/                                 # mdBook pages
pv audit contracts/*.yaml                          # Audit for issues
pv graph contracts/ --format mermaid               # Verification DAG
pv lean contracts/*.yaml                           # Lean 4 theorem stubs

# batuta — falsification
batuta falsify . --format markdown                 # 108-item checklist
batuta oracle --list                               # Stack components
batuta oracle --local                              # Local workspace status

# pmat — code quality (upstream repos)
pmat tdg baseline create                           # TDG baseline
pmat comply check --strict ../aprender

# ═══════════════════════════════════════════════════════════
# VALIDATION (Makefile)
# ═══════════════════════════════════════════════════════════

make validate          # All validation (YAML + contracts + forjar + Makefile)
make lint              # Lint with bashrs
make eval-validate     # Validate benchmark canonical solutions
make dogfood           # Full 12-section dogfooding suite
make book              # Build mdBook
make help              # Show all targets

knowledge-distillation-kernel-v1

Version: 1.0.0

Knowledge distillation kernel — temperature-scaled KL divergence + cross-entropy

References

• Hinton et al. (2015) Distilling the Knowledge in a Neural Network
• Ba & Caruana (2014) Do Deep Nets Really Need to be Deep?

Dependencies

• softmax-kernel-v1
• cross-entropy-kernel-v1

Dependency Graph

graph LR
    knowledge_distillation_kernel_v1["knowledge-distillation-kernel-v1"] --> softmax_kernel_v1["softmax-kernel-v1"]
    knowledge_distillation_kernel_v1["knowledge-distillation-kernel-v1"] --> cross_entropy_kernel_v1["cross-entropy-kernel-v1"]

Equations

kd_loss

$$ L_KD = alpha * KL(softmax(z_t/T) || softmax(z_s/T)) * T^2 + (1-alpha) * CE(y, z_s) $$

Domain: $z_t, z_s in R^V, T > 0, alpha in [0,1]$

Codomain: $L_KD in [0, +inf)$

Invariants:

• $L_KD >= 0 (non-negativity from KL and CE non-negativity)$
• $alpha=0 => L_KD = CE(y, z_s) (pure hard label)$
• $alpha=1 => L_KD = T^2 * KL(teacher || student) (pure soft label)$

kl_divergence

$$ KL(P || Q) = sum_i P(i) * \log(P(i) / Q(i)) $$

Domain: $P, Q valid probability distributions over V classes$

Codomain: $KL in [0, +inf)$

Invariants:

• $KL(P || Q) >= 0 (Gibbs inequality)$
• $KL(P || P) = 0 (identity)$

temperature_softmax

$$ softmax(z/T)_i = \exp(z_i/T) / sum_j \exp(z_j/T) $$

Domain: $z in R^V, T > 0$

Codomain: $softmax in (0, 1)^V, sum = 1$

Invariants:

• $All outputs strictly positive$
• $Outputs sum to 1$
• $T -> inf => uniform distribution$
• $T -> 0 => one-hot on argmax$

Proof Obligations

Kernel Phases

1. teacher_softmax: Compute softmax(z_t / T) — teacher soft targets — output is valid probability distribution
2. student_softmax: Compute softmax(z_s / T) — student soft predictions — output is valid probability distribution
3. kl_divergence: Compute KL(teacher || student) — result >= 0
4. cross_entropy: Compute CE(y, z_s) — hard label loss — result >= 0
5. combine: Combine: alpha * T^2 * KL + (1-alpha) * CE — result >= 0

Falsification Tests

Kani Harnesses

QA Gate

Knowledge Distillation Contract (F-KD-001)

KD loss correctness for Albor distillation pipeline

Checks: kl_non_negativity, temperature_validity, alpha_interpolation, gradient_correctness

Pass criteria: All 5 falsification tests pass + 2 Kani harnesses verify

bpe-tokenizer-kernel-v1

Version: 1.0.0

BPE tokenizer kernel — byte-pair encoding with lossless roundtrip

References

• Sennrich et al. (2016) Neural Machine Translation of Rare Words with Subword Units
• Gage (1994) A New Algorithm for Data Compression

Equations

bpe_merge

$$ merge(a, b) = ab where (a,b) = argmin_{(p,q) in pairs} rank(p,q) $$

Domain: $token sequence with adjacent pairs$

Codomain: $shorter token sequence$

Invariants:

• $Each merge reduces sequence length by at least 1$
• $Merge ordering is deterministic$
• $Final sequence uses only tokens in vocabulary$

roundtrip

$$ decode(encode(x)) = x for all x in UTF-8 $$

Domain: $x: valid UTF-8 string$

Codomain: $encode(x): Vec where each id in [0, V)$

Invariants:

• $Lossless roundtrip for all valid UTF-8$
• $Empty input maps to empty output$
• $Byte-level fallback ensures all byte values representable$

Proof Obligations

Kernel Phases

1. byte_encode: Convert UTF-8 string to byte sequence — bytes are valid UTF-8 representation
2. initial_tokenize: Map bytes to initial token IDs (byte-level) — all bytes have a token mapping
3. bpe_merge: Iteratively apply BPE merge rules in priority order — sequence length decreases monotonically
4. output: Return final token ID sequence — all IDs in [0, vocab_size)

Falsification Tests

Kani Harnesses

QA Gate

BPE Tokenizer Contract (F-TOK-001)

Tokenizer correctness for Albor vocabulary

Checks: roundtrip_lossless, byte_completeness, deterministic_encoding, fim_sentinel_atomic

Pass criteria: All 4 falsification tests pass + Kani roundtrip harness verifies

gradient-accumulation-kernel-v1

Version: 1.0.0

Gradient accumulation kernel — numerical equivalence of micro-batch accumulation

References

• Goyal et al. (2017) Accurate, Large Minibatch SGD: Training ImageNet in 1 Hour

Dependencies

• adamw-kernel-v1

Dependency Graph

graph LR
    gradient_accumulation_kernel_v1["gradient-accumulation-kernel-v1"] --> adamw_kernel_v1["adamw-kernel-v1"]

Equations

accumulation

$$ G_accum = (1/N) * sum_{i=1}^{N} g_i $$

Domain: $g_i: gradient from micro-batch i, N: accumulation steps$

Codomain: $G_accum: accumulated gradient tensor$

Invariants:

• $G_accum approximates G_full within fp tolerance$
• $N=1 => G_accum = g_1 exactly$

loss_scaling

$$ L_scaled = (1/N) * L_micro $$

Domain: $L_micro: micro-batch loss, N: accumulation steps$

Codomain: $L_scaled: scaled loss for backward pass$

Invariants:

• $Total loss = mean of micro-batch losses (not sum)$
• $Gradients are correctly scaled by 1/N$

Proof Obligations

Kernel Phases

1. zero_gradients: Zero gradient buffers at start of accumulation cycle — all gradient values are 0.0
2. accumulate: Add scaled micro-batch gradients: G += (1/N) * g_i — accumulation buffer is fp32
3. clip: Apply gradient clipping to accumulated gradient — ||G_clipped|| <= max_norm
4. step: Optimizer updates parameters using accumulated gradient — called exactly once per N micro-batches

Falsification Tests

Kani Harnesses

QA Gate

Gradient Accumulation Contract (F-GA-001)

Gradient accumulation correctness for Albor training

Checks: numerical_equivalence, gradient_zeroing, step_count, clip_after_accumulate

Pass criteria: All 4 falsification tests pass + 2 Kani harnesses verify

model-merging-kernel-v1

Version: 1.0.0

Model merging kernel — SLERP, TIES, and DARE weight interpolation

References

• Shoemake (1985) Animating Rotation with Quaternion Curves (SLERP)
• Yadav et al. (2023) TIES-Merging: Resolving Interference When Merging Models
• Yu et al. (2023) Language Models are Super Mario: Absorbing Abilities from Homologous Models as a Free Lunch (DARE)

Equations

dare

$$ tau_tilde_i = m_i * tau_i / (1-p) where m_i ~ Bernoulli(1-p) $$

Domain: $tau_i (task vector), p in [0, 1) (drop probability)$

Codomain: $tau_tilde_i: rescaled sparse task vector$

Invariants:

• $E[tau_tilde] = tau (unbiased estimator)$
• $Sparsity approximately p$

slerp

$$ SLERP(w1, w2, t) = sin((1-t)Omega)/sin(Omega) * w1 + sin(tOmega)/sin(Omega) * w2 $$

Domain: $w1, w2 in R^n (weight vectors), t in [0, 1], cos(Omega) = w1.w2 / (||w1|| * ||w2||)$

Codomain: $result in R^n with ||result|| approximately ||w1||$

Invariants:

• $SLERP(w1, w2, 0) = w1 (left boundary)$
• $SLERP(w1, w2, 1) = w2 (right boundary)$
• $||SLERP(w1, w2, t)|| approximately ||w1|| for normalized inputs$

ties

$$ w_merged = w_base + lambda * elect(trim(tau_1, …, tau_n)) $$

Domain: $tau_i = w_i - w_base (task vectors), trim ratio k in [0,1]$

Codomain: $w_merged in R^n$

Invariants:

• $After trim(k%), exactly k% of delta weights are zeroed per layer$
• $Sign election resolves conflicts by majority vote$

Proof Obligations

Kernel Phases

1. validate_architectures: Verify all input models have same architecture — hidden_size, num_layers, vocab_size match
2. compute_task_vectors: Compute delta from base: tau_i = w_i - w_base — tau has same shape as w
3. merge_weights: Apply SLERP/TIES/DARE to combine weights — output weights are finite

Falsification Tests

Kani Harnesses

QA Gate

Model Merging Contract (F-MERGE-001)

Weight merging correctness for Albor post-training

Checks: slerp_bound, slerp_boundary, dare_unbiased

Pass criteria: All 3 falsification tests pass + Kani SLERP harness verifies

pruning-kernel-v1

Version: 1.0.0

Pruning kernel — WANDA and magnitude-based weight pruning

References

• Sun et al. (2023) A Simple and Effective Pruning Approach for Large Language Models (WANDA)
• Han et al. (2015) Learning both Weights and Connections for Efficient Neural Networks

Equations

magnitude_score

$$ score(w_ij) = |w_ij| $$

Domain: $w_ij: weight value$

Codomain: $score in [0, +inf)$

Invariants:

• $score >= 0$
• $score = 0 iff w_ij = 0$

sparsity

$$ s = |{w : w = 0}| / |w| $$

Domain: $w: weight tensor$

Codomain: $s in [0, 1]$

Invariants:

• $s = 0 means no pruning$
• $s = 1 means all weights zeroed$
• $After pruning with target s, achieved sparsity within 0.1% of s$

wanda_score

$$ score(w_ij) = |w_ij| * ||X_j||_2 $$

Domain: $w_ij: weight, X_j: activation column vector$

Codomain: $score in [0, +inf)$

Invariants:

• $score >= 0 (product of norms)$
• $score = 0 iff w_ij = 0 or X_j = 0$

Proof Obligations

Kernel Phases

1. compute_scores: Compute importance score for each weight — scores are non-negative
2. determine_threshold: Find threshold score for target sparsity — threshold partitions weights into keep/prune sets
3. apply_mask: Zero out weights below threshold — sparsity matches target within tolerance

Falsification Tests

Kani Harnesses

QA Gate

Pruning Contract (F-PRUNE-001)

Weight pruning correctness for Albor model compression

Checks: sparsity_guarantee, score_ordering, identity_at_zero

Pass criteria: All 3 falsification tests pass + Kani sparsity harness verifies

training-memory-kernel-v1

Version: 1.0.0

Training memory estimation kernel — closed-form VRAM projection from architecture

References

• Korthikanti et al. (2022) Reducing Activation Recomputation in Large Transformer Models
• Rajbhandari et al. (2020) ZeRO: Memory Optimizations Toward Training Trillion Parameter Models

Dependency Graph

graph LR
    training_gpu_kernel_v1["training-gpu-kernel-v1"] --> training_memory_kernel_v1["training-memory-kernel-v1"]

Equations

activation_memory

$$ M_act = L × S × H × K × 4 K = 10 (Q, K, V, attn_scores, attn_out, gate, up, down, 2×residual)

$$

Domain: $L: num_layers, S: seq_len, H: hidden_size, K: activation tensor count per layer (upper bound), 4: bytes per f32 element $

Codomain: $M_act: peak activation memory in bytes (upper bound)$

Invariants:

• $entrenar processes batch items sequentially — activation memory is per single sequence$
• $K=10 is conservative upper bound; actual depends on tensor lifetime overlap$
• $Gradient checkpointing reduces M_act to O(\sqrt{L}) but is not default$

gradient_memory

$$ M_grad = P_total × 4 $$

Domain: $P_total: parameter count$

Codomain: $M_grad: gradient memory in bytes (exact)$

Invariants:

• $Gradients always f32 regardless of mixed precision mode$
• $One gradient tensor per parameter$

optimizer_memory

$$ M_opt = P_total × 8 $$

Domain: $P_total: parameter count$

Codomain: $M_opt: AdamW optimizer state memory in bytes (exact)$

Invariants:

• $AdamW stores first moment (m) and second moment (v), both f32$
• $M_opt = P × 4 (m) + P × 4 (v) = P × 8$

parameter_count

$$ P_embed = V × H P_layer = 2H + H² + H×D_kv + H×D_kv + H² + H×I + H×I + I×H = 2H + 2H² + 2H×D_kv + 3H×I P_norm = H P_total = P_embed + L × P_layer + P_norm

$$

Domain: $V: vocab_size, H: hidden_size, L: num_hidden_layers, D_kv: num_kv_heads × head_dim, I: intermediate_size, head_dim: H / num_attention_heads $

Codomain: $P_total: total trainable parameter count (exact)$

Invariants:

• $P_total is deterministic given architecture — no randomness$
• $P_embed dominates for large vocab; P_layer dominates for deep models$

total_memory

$$ M_total = M_weights + M_grad + M_opt + M_act + M_cuda $$

Domain: $M_cuda \approx 512 MB (CUDA context, cuBLAS workspace, allocator overhead)$

Codomain: $M_total: total estimated memory in bytes$

Invariants:

• $M_total is an upper bound — actual usage may be lower due to tensor reuse$
• $Does not include KV cache (inference only, not training)$
• $entrenar hybrid mode: weights/grads/optimizer live in CPU RAM; only matmul operands transfer to GPU$
• $In hybrid mode, VRAM \approx M_cuda + max(matmul_operand_pair); CPU RAM \approx M_weights + M_grad + M_opt + M_act$
• $M_total represents peak system memory (CPU+GPU) needed, not VRAM alone$

weight_memory

$$ M_weights = P_total × B_w $$

Domain: $P_total: parameter count, B_w: bytes per weight (4 for f32, 2 for fp16/bf16)$

Codomain: $M_weights: weight memory in bytes (exact)$

Invariants:

• $Mixed precision stores master weights in f32 + fp16 copy: M_weights = P × (4 + 2)$
• $entrenar current impl: always f32 storage, fp16 cast at matmul site$

Proof Obligations

Falsification Tests

Kani Harnesses

QA Gate

Training Memory Estimation Contract (F-MEM-001)

VRAM estimation correctness for apr train plan

Checks: parameter_count_exact, activation_upper_bound, total_covers_actual

Pass criteria: All 3 falsification tests pass

training-gpu-kernel-v1

Version: 1.0.0

GPU-resident pretraining kernel — CudaTransformerBlock wired into TransformerTrainer

References

• classify_pipeline.rs GPU training pattern (ENT-151, ENT-152)
• training-memory-kernel-v1.yaml (VRAM estimation)

Dependencies

• training-memory-kernel-v1

Dependency Graph

graph LR
    training_gpu_kernel_v1["training-gpu-kernel-v1"] --> training_memory_kernel_v1["training-memory-kernel-v1"]

Equations

gpu_utilization

$$ util = compute_time / (compute_time + transfer_time + sync_time)

$$

Domain: $Measured via nvidia-smi dmon or CUDA events$

Codomain: $GPU utilization ratio [0, 1]$

Invariants:

• $util > 0.70 for models >= 350M params with batch_size >= 4$
• $Previous CPU autograd achieved ~0.07 (7%) due to 16K transfers/step$

pcie_transfers_per_step

$$ T = 3 (constant) Transfer 1 (H2D): hidden = S × H × 4 bytes Transfer 2 (D2H): logits = S × V × 4 bytes Transfer 3 (H2D): grad_logits = S × V × 4 bytes Total bytes per step = S × (H + 2V) × 4

$$

Domain: $S: seq_len, H: hidden_size, V: vocab_size $

Codomain: $T = 3: exactly 3 PCIe transfers per training step$

Invariants:

• $Embedding lookup stays on CPU (scatter-gather, not matmul)$
• $Cross-entropy loss + softmax backward stays on CPU$
• $All transformer block forward/backward/optimizer on GPU$
• $RMSNorm forward/backward on GPU$
• $LM head GEMM forward/backward on GPU$

transfer_overhead

$$ overhead_ms = total_bytes / bandwidth For PCIe 4.0 x16: bandwidth = 32 GB/s For 350M model (H=1024, V=32K, S=2048): total = 2048 × (1024 + 2×32768) × 4 = 544 MB overhead = 544 MB / 32 GB/s = 17 ms

$$

Domain: $Architecture params + PCIe bandwidth$

Codomain: $Transfer overhead in milliseconds (theoretical)$

Invariants:

• $Transfer overhead < 5% of compute time for models >= 350M params$
• $GPU compute time dominates for large models$

Proof Obligations

Falsification Tests

QA Gate

GPU-Resident Pretraining Contract (F-GPU-001)

CudaTransformerTrainer correctness and efficiency

Checks: numerical_equivalence, transfer_count_invariant, gpu_utilization_bound, weight_sync_exact, graceful_fallback

Pass criteria: All 4 falsification tests pass

Training Step Budget Contract

Contract: contracts/training-step-budget-v1.yaml Version: 1.0.0 Status: NEW (ALB-075) Depends on: training-gpu-kernel-v1, cublas-gemm-v1

Equations

step_time_budget

T_step = T_gemm + T_optimizer + T_embedding + T_pcie + T_elementwise
       + T_cross_entropy + T_stream_sync + T_overhead

Every component maps to exactly one probador brick. Budget violation (> 2x) triggers Jidoka alert.

gemm_throughput

TFLOP_per_step = sum(2 * m * n * k / 1e12 for all ~555 GEMMs)
T_gemm = TFLOP_per_step / achieved_tflops

• PTX baseline: ~2 TFLOP/s
• cuBLAS target: >= 100 TFLOP/s

mfu_definition

MFU = (6 * P * tokens_per_step) / (T_step * peak_flops)
P = 370M, tokens_per_step = 4096
peak_flops(FP16, sustained) = 148 TFLOP/s

Proof Obligations (4)

Falsification Tests (4)

QA Gate

F-BUDGET-001: All 4 falsification tests must pass before optimization phase targets are considered valid.

cuBLAS GEMM Integration Contract

Contract: contracts/cublas-gemm-v1.yaml Version: 1.0.0 Status: NEW (ALB-075) Depends on: training-gpu-kernel-v1, training-memory-kernel-v1

Equations

cublas_gemm_correctness

C_cublas = alpha * op(A) * op(B) + beta * C
where op(X) = X if transa=N, X^T if transa=T
A: FP16 [m, k], B: FP16 [k, n], C: FP16 [m, n]
Accumulation: FP32 (CUBLAS_COMPUTE_32F)

• max_abs_diff(C_cublas, C_ptx) < 1e-2 for identical inputs
• cuBLAS uses tensor cores when math mode is TENSOR_OP_MATH
• FP32 accumulation prevents catastrophic cancellation

buffer_size_verification

For cublasGemmEx(m, n, k, A, B, C):
  A.len() >= m * k * 2  (FP16)
  B.len() >= k * n * 2  (FP16)
  C.len() >= m * n * 2  (FP16)

Verified at call site, not inside cuBLAS. Assertion failure = immediate panic.

handle_lifecycle

create: cublasCreate_v2(&handle) -> CUBLAS_STATUS_SUCCESS
bind:   cublasSetStream_v2(handle, stream) once per training step
drop:   cublasDestroy_v2(handle) exactly once

• One handle per CudaContext (thread-safe within context)
• Stream set ONCE per step, not per GEMM (555 calls = measurable overhead)
• Handle destroyed on Drop (Rust RAII)

ffi_overhead

overhead = T_rust_cublas / T_raw_c_cublas < 1.02

For identical GEMM shape, same GPU, same cuBLAS config. Measured via CUDA events, not wall clock. Warmup: 50 iterations discarded before measurement.

mfu_improvement

MFU = (6 * P * tokens_per_step) / (T_step * peak_flops)
P = 370M, tokens_per_step = 4096
peak_flops(FP16, sustained) = 148 TFLOP/s

• MFU(cublas) > MFU(ptx) (strict improvement)
• MFU(cublas) >= 0.025 (must beat current 2.5% FP32 baseline)

mixed_precision_weight_flow

CPU master weights: FP32 (optimizer operates here)
GPU forward weights: FP16 (cast during upload)
GPU activation gradients: FP16 (cuBLAS backward output)
GPU weight gradients: FP32 (accumulated in FP32 buffer)
CPU gradient download: FP32 (for optimizer update)

• Master weights ALWAYS FP32 on CPU (no precision loss in optimizer)
• C-EMBED-GRAD-001 still holds: activation grad clipped before CPU scatter-add
• C-HYPERPARAMS-001 still holds: all optimizer params from YAML config

Proof Obligations (8)

Falsification Tests (11)

Kani Harness

KANI-CUBLAS-001: Buffer size assertion prevents overflow for all valid GEMM shapes (exhaustive, bound=8).

QA Gate

F-CUBLAS-001: All 11 falsification tests must pass before cuBLAS backend replaces PTX for training.

Fused Kernel Optimizations Contract

Contract: contracts/fused-kernels-v1.yaml Version: 1.0.0 Status: NEW (ALB-075 Phase 4+) Depends on: cublas-gemm-v1, training-gpu-kernel-v1, training-step-budget-v1 Source: unslothai/unsloth analysis

Equations

fused_cross_entropy

For each row r in logits [B*S, V]:
  logsumexp_r = log(sum(exp(logit[r, i])))
  loss_r = logsumexp_r - logit[r, label_r]
  grad_r[i] = exp(logit[r, i] - logsumexp_r) - delta(i, label_r)

Single kernel pass. FP32 accumulation. Softmax tensor never materialized. Backward grad overwrites logits buffer in-place (zero extra allocation).

rmsnorm_activation_reuse

Forward: save ONLY inv_var [B*S] (not normed — recompute in backward)
Backward: normed = X_cached * inv_var_saved (bit-exact recompute)
Memory savings: 24 layers * B * S * H * 4 bytes = 384 MB

swiglu_inplace_backward

d_up = grad_output * silu(gate)          → written to up buffer
d_gate = grad_output * up * silu'(gate)  → written to gate buffer

gate and up consumed before overwrite. Peak workspace reduced by 128 MB.

rope_head_grouping

Load sin/cos once per group (G=4 heads)
Apply to all heads in group with single memory load
Q: 4 groups of 4, K: 1 group of 4

Bit-exact with per-head RoPE. ~10% attention speedup from L2 cache reuse.

fused_tiled_attention

For tile_q, tile_k in tiled [0, S):
  scores_tile = Q[tile_q] @ K[tile_k]^T / sqrt(d_k)
  Online softmax (Milakov & Gimelshein 2018):
    m_new = max(m_old, max(scores_tile))
    l_new = l_old * exp(m_old - m_new) + sum(exp(scores_tile - m_new))
  O += exp(scores_tile - m_new) @ V[tile_k]
O = O / l_new

Full [S, S] attention matrix never materialized. Memory: O(BHSd_k) instead of O(BHSS). Saves 256 MB per layer.

chunked_cross_entropy (deferred)

For vocab > 65K: split logsumexp into 65K chunks. Mathematically exact (logsumexp is associative). Current vocab=32K: single chunk, no overhead.

Proof Obligations (10)

Falsification Tests (10)

Priority Matrix

QA Gate

F-FUSED-001: All 10 falsification tests must pass. If combined run shows instability, bisect fusions individually to identify the culprit.

Training Performance Specification

0. Design Principles

This specification follows design by contract (DbC). Every performance claim, optimization target, and implementation phase begins with a provable contract (pv validate) that defines equations, invariants, proof obligations, and falsification tests. Code is written to satisfy the contract — never the reverse.

Verification stack (sovereign, no external dependencies):

Workflow for every optimization phase:

1. pv validate contracts/cublas-gemm-v1.yaml          # Contract first
2. pv scaffold contracts/cublas-gemm-v1.yaml           # Generate test stubs
3. make bench-gemm-raw                                 # Establish ceiling
4. Implement against contract
5. make bench-gemm-compare                             # Three-tier benchmark
6. probador brick budgets verify per-component SLAs    # Brick profiling
7. renacer --trace-compute traces per-kernel timing    # Layer tracing
8. pv audit contracts/cublas-gemm-v1.yaml              # Binding coverage
9. Dogfood on 350M training run
10. make bench-gemm-regression                         # No regressions
11. Close gap in §11

1. Current Performance Baseline

1.1 Measured Throughput

1.2 MFU Analysis

Model FLOPs Utilization (MFU) measures actual compute throughput against hardware theoretical peak. For a transformer forward+backward pass, the standard approximation is 6 x params x tokens_per_step FLOPs.

Model parameters:       370M (24 layers, hidden=1024, intermediate=4096)
Tokens per step:        4 x 1024 = 4,096 tokens
FLOPs per step:         6 x 370M x 4,096 = 9.1 TFLOP

Step time:              4.4s
Achieved FLOP/s:        9.1 TFLOP / 4.4s = 2.07 TFLOP/s

RTX 4090 FP16 peak:    165 TFLOP/s (with tensor cores)
RTX 4090 FP32 peak:    82.6 TFLOP/s (without tensor cores)

MFU (vs FP16 peak):    2.07 / 165 = 1.3%
MFU (vs FP32 peak):    2.07 / 82.6 = 2.5%

MFU = 2.5% (vs FP32 peak) / 1.3% (vs FP16 peak)

1.3 Research Benchmarks for Context

The gap is 10-15x vs what the hardware can deliver for this model size.

1.4 Baseline Profiling Protocol (renacer + probador)

Before any optimization, establish ground truth with brick-level profiling:

# Layer-level tracing: per-kernel timing for one training step
renacer --otlp-endpoint http://localhost:4317 \
        --otlp-service-name "albor-baseline" \
        --trace-compute \
        --trace-compute-threshold 100 \
        -- apr train apply --task pretrain \
            --config configs/train/pretrain-350m-cuda-test.yaml

# View in Jaeger: http://localhost:16686 -> Service: "albor-baseline"
# Each GEMM kernel, norm kernel, PCIe transfer is a span with duration_us

BrickTracer escalation thresholds for baseline measurement:

#![allow(unused)]
fn main() {
let thresholds = BrickEscalationThresholds::default()
    .with_cv(15.0)         // Escalate if kernel timing CV > 15%
    .with_efficiency(25.0)  // Escalate if compute efficiency < 25%
    .with_rate_limit(100);  // Max 100 traces/second during profiling
}

Brick budget breakdown (probador) — defines the per-component SLA that each optimization phase must improve:

#![allow(unused)]
fn main() {
let step_budget = BrickHouseBuilder::new("training-step")
    .budget_ms(4400)                      // Current step time
    .brick("gemm_forward",     1400)      // 7 GEMMs x 24 blocks + LM head
    .brick("gemm_backward",    1100)      // 14 GEMMs x 24 blocks + LM head
    .brick("cpu_optimizer",     800)      // 24 blocks + LM head + embedding
    .brick("cpu_embedding",     200)      // Scatter-gather forward + backward
    .brick("pcie_transfer",     150)      // 3 transfers (H2D embed, D2H logits, H2D grad)
    .brick("elementwise_kernel", 100)     // RMSNorm, RoPE, SiLU
    .brick("cross_entropy",      50)      // Fused CE forward + backward
    .brick("stream_sync",        50)      // ALB-065 synchronization
    .brick("overhead",          550)      // Scheduling, allocator, host logic
    .build()?;
}

Each brick has a Jidoka gate: if any component exceeds its budget by >2x after an optimization, training stops and alerts. This prevents silent regressions.

2. Root Cause Analysis

2.1 The GEMM Bottleneck

A 350M transformer forward+backward step executes 552 GEMM operations:

Per transformer block (24 blocks):
  Forward:
    - Q projection:    GEMM [S, H] x [H, H]     (1)
    - K projection:    GEMM [S, H] x [H, H_kv]  (1)
    - V projection:    GEMM [S, H] x [H, H_kv]  (1)
    - Attention out:   GEMM [S, H] x [H, H]     (1)
    - FFN gate:        GEMM [S, H] x [H, I]     (1)
    - FFN up:          GEMM [S, H] x [H, I]     (1)
    - FFN down:        GEMM [S, I] x [I, H]     (1)
  Backward (roughly 2x forward):
    - dQ, dK, dV, dAttn_out, dGate, dUp, dDown  (7)
    - Weight gradients for each of the above     (7)
  Subtotal per block: 7 + 14 = 21 GEMMs

LM head (vocab projection):
  Forward:   GEMM [S, H] x [H, V]               (1)
  Backward:  GEMM for dInput + dWeight           (2)
  Subtotal: 3 GEMMs

Embedding (scatter-add, not GEMM):              (0)

Total: 24 x 21 + 3 = 507 weight GEMMs
       + attention score GEMMs: 24 x 2 = 48 (QK^T forward + backward)
       = 555 GEMM operations per step

2.2 Hand-Written PTX vs Tensor Cores

All GEMMs use hand-written PTX tiled GEMM kernels in trueno-gpu:

• GemmForwardKernel::tiled_unrolled() — FP32 accumulation, no tensor cores
• GemmBackwardAKernel::tiled_unrolled() — Input gradient GEMM
• GemmBackwardBKernel::tiled_unrolled() — Weight gradient GEMM

These kernels:

• Use scalar FP32 FMA instructions (fma.rn.f32)
• Tile sizes are small (typically 16x16 or 32x32)
• No shared memory double-buffering or software pipelining
• Cannot use tensor cores (require wmma or mma PTX instructions)

The RTX 4090 (Ada Lovelace, SM 8.9) has 128 FP32 CUDA cores per SM x 128 SMs = 16,384 CUDA cores. But it also has 4th generation tensor cores that deliver 165 TFLOP/s FP16 — 2x the FP32 throughput — and these are completely unused.

2.3 Non-GEMM Overhead

The per-block CPU optimizer (download gradients -> AdamW on CPU -> upload weights) is the second largest bottleneck after GEMM throughput. ALB-067 disabled per-block gradient clipping due to CPU-side L2 norm cost (864 D2H transfers/step).

2.4 Step Time Breakdown (Estimated)

Total step time:          4,400 ms (100%)
+-- 555 GEMM operations:  2,500 ms ( 57%)  <-- PRIMARY BOTTLENECK
+-- CPU optimizer (24x):    800 ms ( 18%)  <-- SECONDARY BOTTLENECK
+-- CPU embedding:          200 ms (  5%)
+-- PCIe transfers:         150 ms (  3%)
+-- Element-wise kernels:   100 ms (  2%)
+-- Cross-entropy:           50 ms (  1%)
+-- Stream sync:             50 ms (  1%)
+-- Overhead (Python-free):  550 ms ( 13%)

2.5 Confirming the Breakdown: Layer Tracing Protocol

The estimated breakdown in 2.4 must be confirmed with measurement before optimizing. Renacer BrickTracer provides per-brick isolation:

#![allow(unused)]
fn main() {
// In entrenar CudaTransformerTrainer::train_step_single()
let tracer = BrickTracer::new_local();

// Trace each phase as a separate brick
let embed_result = tracer.trace("embed_forward", 200, || {
    // CPU scatter-gather embedding lookup
    embed_forward(&input_ids, &embed_weight)
});

let h2d_result = tracer.trace("pcie_h2d_hidden", 50, || {
    hidden_buf.copy_from_host(&hidden_states)
});

for block_idx in 0..24 {
    let fwd_result = tracer.trace(
        &format!("block_{}_forward", block_idx), 100, || {
            block.forward(&workspace)
        }
    );
    // BrickTracer records: duration_us, budget_us, efficiency, over_budget
}
}

Escalation: When any brick’s CV exceeds 15% (unstable timing) or efficiency drops below 25% (idle GPU), BrickTracer automatically captures full syscall-level traces and exports as OTLP spans. This is the renacer “measurement -> tracing” escalation pattern — lightweight metrics in steady state, detailed tracing only on anomaly.

The confirmed breakdown becomes the contract baseline that optimization phases are proven against.

3. Contracts: Write Before Code

3.1 Contract: cuBLAS GEMM Integration

File: contracts/cublas-gemm-v1.yaml

This contract must be written and validated (pv validate) before any cuBLAS code is written. It defines the algebraic invariants, numerical bounds, and falsification tests that the implementation must satisfy.

# contracts/cublas-gemm-v1.yaml
metadata:
  version: "1.0.0"
  created: "2026-03-05"
  author: "PAIML Engineering"
  description: "cuBLAS tensor core GEMM integration for training throughput"
  references:
    - "Micikevicius et al. (2018) Mixed Precision Training"
    - "NVIDIA cuBLAS Documentation (CUDA 12.x)"
    - "training-gpu-kernel-v1.yaml (parent contract)"
  depends_on:
    - "training-gpu-kernel-v1"
    - "training-memory-kernel-v1"

equations:
  cublas_gemm_correctness:
    formula: |
      C_cublas = alpha * op(A) * op(B) + beta * C
      where op(X) = X if transa=N, X^T if transa=T
      A: FP16 [m, k], B: FP16 [k, n], C: FP16 [m, n]
      Accumulation: FP32 (CUBLAS_COMPUTE_32F)
    domain: "FP16 input buffers, FP32 accumulation, FP16 output"
    codomain: "C_cublas: FP16 result matrix"
    invariants:
      - "max_abs_diff(C_cublas, C_ptx) < 1e-2 for identical inputs"
      - "cuBLAS uses tensor cores when math mode is TENSOR_OP_MATH"
      - "FP32 accumulation prevents catastrophic cancellation"

  buffer_size_verification:
    formula: |
      For cublasGemmEx(m, n, k, A, B, C):
        A.len() >= m * k * sizeof(FP16) = m * k * 2
        B.len() >= k * n * sizeof(FP16) = k * n * 2
        C.len() >= m * n * sizeof(FP16) = m * n * 2
    domain: "GpuBuffer lengths in bytes"
    codomain: "Boolean: all buffers sufficient"
    invariants:
      - "Verified at call site, not inside cuBLAS (Rule 2: prove at kernel boundary)"
      - "Assertion failure = immediate panic, not silent corruption"

  handle_lifecycle:
    formula: |
      create: cublasCreate_v2(&handle) -> CUBLAS_STATUS_SUCCESS
      bind:   cublasSetStream_v2(handle, stream) before every GEMM
      drop:   cublasDestroy_v2(handle) exactly once
    invariants:
      - "One handle per CudaContext (thread-safe within context)"
      - "Stream set before EVERY cublasGemmEx call (C-STREAMSYNC-001 extension)"
      - "Handle destroyed on Drop (Rust RAII)"
      - "No default stream usage — always explicit non-blocking stream"

  mfu_improvement:
    formula: |
      MFU = achieved_flops / hardware_peak_flops
      achieved_flops = 6 * P * tokens_per_step / step_time
      P = 370M, tokens_per_step = 4096
      hardware_peak_flops(FP16) = 165 TFLOP/s
    domain: "Measured step_time after cuBLAS integration"
    codomain: "MFU ratio [0, 1]"
    invariants:
      - "MFU(cublas) > MFU(ptx) (strict improvement)"
      - "MFU(cublas) >= 0.025 (must beat current 2.5% FP32 baseline)"

  mixed_precision_weight_flow:
    formula: |
      CPU master weights: FP32 (optimizer operates here)
      GPU forward weights: FP16 (cast during upload)
      GPU activation gradients: FP16 (cuBLAS backward output)
      GPU weight gradients: FP32 (accumulated in FP32 buffer)
      CPU gradient download: FP32 (for optimizer update)
    invariants:
      - "Master weights ALWAYS FP32 on CPU (no precision loss in optimizer)"
      - "Weight gradient accumulation in FP32 (no underflow in small gradients)"
      - "C-EMBED-GRAD-001 still holds: activation grad clipped before CPU scatter-add"
      - "C-HYPERPARAMS-001 still holds: all optimizer params from YAML config"

proof_obligations:
  - type: equivalence
    property: "cuBLAS GEMM matches PTX GEMM"
    formal: "max_abs_diff(C_cublas, C_ptx) < 1e-2 for all GEMM shapes in training"
    tolerance: 1e-2
    applies_to: cublas_gemm_correctness

  - type: invariant
    property: "Buffer sizes verified before every cublasGemmEx"
    formal: "assert!(buf.len() >= required) precedes every cublasGemmEx call"
    tolerance: 0
    applies_to: buffer_size_verification

  - type: invariant
    property: "cuBLAS handle lifecycle is RAII"
    formal: "create() in new(), destroy() in Drop, set_stream() before gemm()"
    tolerance: 0
    applies_to: handle_lifecycle

  - type: bound
    property: "MFU improves over baseline"
    formal: "MFU(cublas, 50 steps) > MFU(ptx, 50 steps)"
    applies_to: mfu_improvement

  - type: invariant
    property: "Training stability preserved"
    formal: "loss.is_finite() for all steps in 100-step run"
    tolerance: 0
    applies_to: training_stability

  - type: invariant
    property: "Gradient flow preserved"
    formal: "max(|grad(param)|) > 0 for all trainable params after 1 step"
    tolerance: 0
    applies_to: gradient_flow

  - type: invariant
    property: "FP32 accumulation enforced"
    formal: "computeType == CUBLAS_COMPUTE_32F for every cublasGemmEx call"
    tolerance: 0
    applies_to: cublas_gemm_correctness

falsification_tests:
  - id: FALSIFY-CUBLAS-001
    rule: "cuBLAS forward matches PTX forward"
    prediction: "max_abs_diff(logits_cublas, logits_ptx) < 1e-2 on 50M model"
    test: |
      Build TransformerConfig::tiny(), forward same input through both backends.
      Compare logit tensors element-wise.
    if_fails: "cuBLAS transpose convention or leading dimension wrong"

  - id: FALSIFY-CUBLAS-002
    rule: "cuBLAS training stable for 50 steps"
    prediction: "Loss is finite at every step, loss curve within 5% of PTX baseline"
    test: |
      Train 50M model for 50 steps with cuBLAS backend.
      Train same model for 50 steps with PTX backend.
      Compare loss at step 50: |loss_cublas - loss_ptx| / loss_ptx < 0.05.
    if_fails: "FP16 precision insufficient for this model or gradient accumulation broken"

  - id: FALSIFY-CUBLAS-003
    rule: "GEMM throughput exceeds 100 TFLOP/s"
    prediction: "Isolated GEMM [4096, 1024] x [1024, 4096] > 100 TFLOP/s"
    test: |
      Run 1000 iterations of cublasGemmEx on [4096, 1024] x [1024, 4096].
      Compute FLOP/s = 2 * 4096 * 1024 * 4096 * 1000 / elapsed_seconds.
    if_fails: "Tensor cores not engaged, wrong math mode, or memory bandwidth bound"

  - id: FALSIFY-CUBLAS-004
    rule: "Step time improves over PTX baseline"
    prediction: "350M step time < 3.0s with cuBLAS (vs 4.4s with PTX)"
    test: |
      Run pretrain-350m-cuda-test.yaml for 50 steps with cuBLAS.
      Measure median step time. Must be < 3.0s.
    if_fails: "GEMM is not the bottleneck or cuBLAS adds unexpected overhead"

  - id: FALSIFY-CUBLAS-005
    rule: "Buffer overflow impossible"
    prediction: "cuBLAS wrapper panics if buffer too small (never silent corruption)"
    test: |
      Call gemm_f16() with undersized C buffer (m*n*2 - 1 bytes).
      Must panic with assertion failure, not proceed to cublasGemmEx.
    if_fails: "Buffer verification missing or assertion not checked"

  - id: FALSIFY-CUBLAS-006
    rule: "All trainable parameters receive gradients"
    prediction: "max(|grad|) > 0 for every param after 1 cuBLAS training step"
    test: |
      Train 50M model for 1 step with cuBLAS. Check gradient of all 110 params.
    if_fails: "cuBLAS backward produces zero gradients (wrong transpose or alpha/beta)"

  - id: FALSIFY-CUBLAS-007
    rule: "C-EMBED-GRAD-001 preserved under cuBLAS"
    prediction: "Activation gradient clipped before CPU scatter-add even with cuBLAS"
    test: |
      Train 24-layer 350M for 1 step with cuBLAS. Verify activation gradient
      L2 norm <= max_grad_norm before embedding backward.
    if_fails: "cuBLAS backward bypasses activation gradient clipping path"

kani_harnesses:
  - id: KANI-CUBLAS-001
    obligation: CUBLAS-INV-002
    property: "Buffer size assertion prevents overflow for all valid GEMM shapes"
    bound: 8
    strategy: exhaustive
    harness: verify_buffer_assertion_complete

qa_gate:
  id: F-CUBLAS-001
  name: "cuBLAS GEMM Integration Contract"
  description: "Correctness, stability, performance, and safety for cuBLAS tensor core GEMMs"
  checks:
    - "cublas_gemm_correctness"
    - "buffer_size_verification"
    - "handle_lifecycle"
    - "mfu_improvement"
    - "training_stability"
    - "gradient_flow"
  pass_criteria: "All 7 falsification tests pass"
  falsification: "Use wrong transpose to detect GEMM shape errors (ALB-059 class)"

3.2 Contract: Training Step Performance Budget

File: contracts/training-step-budget-v1.yaml

This contract defines the per-brick performance budget that probador enforces.

# contracts/training-step-budget-v1.yaml
metadata:
  version: "1.0.0"
  created: "2026-03-05"
  author: "PAIML Engineering"
  description: "Training step performance budget — brick-level SLAs with Jidoka gates"
  references:
    - "training-gpu-kernel-v1.yaml"
    - "ALB-067: CPU-side gradient clipping bottleneck"
  depends_on:
    - "training-gpu-kernel-v1"
    - "cublas-gemm-v1"

equations:
  step_time_budget:
    formula: |
      T_step = T_gemm + T_optimizer + T_embedding + T_pcie + T_elementwise
             + T_cross_entropy + T_stream_sync + T_overhead
    domain: "Per-component timing measured by renacer BrickTracer"
    codomain: "T_step: total step time in milliseconds"
    invariants:
      - "T_step is sum of brick times (no unaccounted gaps > 5% of total)"
      - "Every component maps to exactly one probador brick"
      - "Brick budget violation triggers Jidoka alert (training pause)"

  gemm_throughput:
    formula: |
      TFLOP_per_gemm(m, n, k) = 2 * m * n * k / 1e12
      TFLOP_per_step = sum(TFLOP_per_gemm for all 555 GEMMs)
      T_gemm = TFLOP_per_step / achieved_tflops
    invariants:
      - "PTX baseline: achieved_tflops ~= 2 TFLOP/s (FP32 scalar)"
      - "cuBLAS target: achieved_tflops >= 100 TFLOP/s (FP16 tensor core)"

  mfu_definition:
    formula: |
      MFU = (6 * P * tokens_per_step) / (T_step * peak_flops)
      P = 370M, tokens_per_step = batch * seq_len = 4096
      peak_flops(FP16) = 165 TFLOP/s, peak_flops(FP32) = 82.6 TFLOP/s
    invariants:
      - "MFU is measured over >= 50 steps (warm cache, excluding first 5)"
      - "Report both FP16 and FP32 MFU for clarity"

proof_obligations:
  - type: bound
    property: "Brick budgets account for full step time"
    formal: "sum(brick_budgets) >= 0.95 * T_step_measured"
    applies_to: step_time_budget

  - type: bound
    property: "GEMM brick dominates baseline"
    formal: "T_gemm / T_step > 0.50 in PTX baseline"
    applies_to: gemm_throughput

  - type: bound
    property: "cuBLAS reduces GEMM brick time by >= 5x"
    formal: "T_gemm(cublas) < T_gemm(ptx) / 5"
    applies_to: gemm_throughput

  - type: bound
    property: "MFU improves monotonically across phases"
    formal: "MFU(phase_N+1) > MFU(phase_N) for each optimization phase"
    applies_to: mfu_definition

falsification_tests:
  - id: FALSIFY-BUDGET-001
    rule: "Brick budgets cover >= 95% of step time"
    prediction: "T_step - sum(bricks) < 0.05 * T_step"
    test: |
      Run 50-step profiling with BrickTracer on 350M model.
      Sum all brick durations. Compare to total step time.
    if_fails: "Unaccounted overhead — missing brick or hidden synchronization"

  - id: FALSIFY-BUDGET-002
    rule: "GEMM is the primary bottleneck in PTX baseline"
    prediction: "T_gemm > 50% of T_step in PTX mode"
    test: |
      Profile 50 steps with PTX backend, isolate GEMM brick time.
    if_fails: "Bottleneck is elsewhere — revisit optimization target"

  - id: FALSIFY-BUDGET-003
    rule: "Jidoka gate fires on 2x budget violation"
    prediction: "If T_gemm > 2 * budget_gemm, training pauses with alert"
    test: |
      Inject artificial 10s delay in GEMM kernel. Verify Jidoka gate
      fires and training loop emits Andon alert.
    if_fails: "Budget enforcement not wired into training loop"

qa_gate:
  id: F-BUDGET-001
  name: "Training Step Performance Budget Contract"
  checks:
    - "brick_coverage"
    - "gemm_dominance"
    - "jidoka_enforcement"
  pass_criteria: "All 3 falsification tests pass"

3.3 Contract Validation Workflow

# Validate both contracts before writing any code
pv validate contracts/cublas-gemm-v1.yaml
pv validate contracts/training-step-budget-v1.yaml

# Generate test scaffolding
pv scaffold contracts/cublas-gemm-v1.yaml -o trueno-gpu/tests/
pv scaffold contracts/training-step-budget-v1.yaml -o entrenar/tests/

# After implementation: audit binding coverage
pv audit contracts/cublas-gemm-v1.yaml \
    --binding contracts/trueno-gpu/cublas-binding.yaml

# After dogfooding: close gaps
pv audit contracts/training-step-budget-v1.yaml \
    --binding contracts/entrenar/step-budget-binding.yaml

4. cuBLAS Integration Plan

4.1 Why cuBLAS

cuBLAS is NVIDIA’s production GEMM library. It:

• Uses tensor cores automatically (FP16 input -> FP32 accumulate -> FP16 output)
• Has auto-tuned kernels for every GPU architecture since Volta
• Handles tiling, shared memory staging, warp scheduling, and epilogue fusion
• Delivers 80-95% of theoretical peak on large matrices

For the Albor GEMM shapes ([4096, 1024] x [1024, 4096] etc.), cuBLAS will use tensor cores, achieving 130-150 TFLOP/s on RTX 4090 vs the current ~2 TFLOP/s from scalar PTX.

4.2 Architecture

The integration lives in trueno-gpu (the CUDA backend crate), adding three new source files:

trueno-gpu/
+-- src/
    +-- cublas_sys.rs     # Raw FFI bindings (unsafe extern "C")
    +-- cublas.rs         # Safe Rust wrapper (CublasHandle, GemmConfig)
    +-- gemm.rs           # Existing hand-written PTX kernels
    +-- ...

4.2.1 cublas_sys.rs — FFI Bindings (~200 lines)

Minimal bindings for the subset of cuBLAS used by training:

#![allow(unused)]
fn main() {
// Core types
type cublasHandle_t = *mut std::ffi::c_void;

#[repr(C)]
enum cublasOperation_t {
    CUBLAS_OP_N = 0,  // No transpose
    CUBLAS_OP_T = 1,  // Transpose
}

#[repr(C)]
enum cublasStatus_t {
    CUBLAS_STATUS_SUCCESS = 0,
    // ... error codes
}

// Core functions
extern "C" {
    fn cublasCreate_v2(handle: *mut cublasHandle_t) -> cublasStatus_t;
    fn cublasDestroy_v2(handle: cublasHandle_t) -> cublasStatus_t;
    fn cublasSetStream_v2(handle: cublasHandle_t, stream: CUstream) -> cublasStatus_t;
    fn cublasSetMathMode(handle: cublasHandle_t, mode: cublasMath_t) -> cublasStatus_t;

    // The workhorse: C = alpha * op(A) * op(B) + beta * C
    fn cublasGemmEx(
        handle: cublasHandle_t,
        transa: cublasOperation_t,
        transb: cublasOperation_t,
        m: i32, n: i32, k: i32,
        alpha: *const f32,
        A: *const std::ffi::c_void, Atype: cudaDataType,
        lda: i32,
        B: *const std::ffi::c_void, Btype: cudaDataType,
        ldb: i32,
        beta: *const f32,
        C: *mut std::ffi::c_void, Ctype: cudaDataType,
        ldc: i32,
        computeType: cublasComputeType_t,
        algo: cublasGemmAlgo_t,
    ) -> cublasStatus_t;
}
}

Link against libcublas.so (ships with CUDA toolkit, already installed for trueno’s PTX compilation):

# trueno-gpu/build.rs
println!("cargo:rustc-link-lib=cublas");
println!("cargo:rustc-link-search=/usr/local/cuda/lib64");

4.2.2 cublas.rs — Safe Wrapper (~300 lines)

#![allow(unused)]
fn main() {
pub struct CublasHandle {
    handle: cublasHandle_t,
}

impl CublasHandle {
    pub fn new() -> Result<Self, CublasError> { ... }

    pub fn set_stream(&self, stream: &CudaStream) -> Result<(), CublasError> { ... }

    /// C = alpha * A x B + beta * C
    /// A: [m, k], B: [k, n], C: [m, n]
    /// Uses FP16 tensor cores with FP32 accumulation
    pub fn gemm_f16(
        &self,
        m: usize, n: usize, k: usize,
        alpha: f32,
        a: &GpuBuffer,  // FP16 [m, k]
        b: &GpuBuffer,  // FP16 [k, n]
        beta: f32,
        c: &mut GpuBuffer,  // FP16 [m, n]
    ) -> Result<(), CublasError> {
        // C-CUBLAS-003: Buffer sizes verified at kernel boundary (Rule 2)
        assert!(a.len() >= m * k * 2, "A buffer too small");
        assert!(b.len() >= k * n * 2, "B buffer too small");
        assert!(c.len() >= m * n * 2, "C buffer too small");

        unsafe {
            check_status(cublasGemmEx(
                self.handle,
                CUBLAS_OP_N, CUBLAS_OP_N,
                m as i32, n as i32, k as i32,
                &alpha,
                a.ptr(), CUDA_R_16F, m as i32,
                b.ptr(), CUDA_R_16F, k as i32,
                &beta,
                c.mut_ptr(), CUDA_R_16F, m as i32,
                CUBLAS_COMPUTE_32F,         // C-CUBLAS-004: FP32 accumulation
                CUBLAS_GEMM_DEFAULT_TENSOR_OP,
            ))
        }
    }
}

impl Drop for CublasHandle {
    fn drop(&mut self) {
        unsafe { cublasDestroy_v2(self.handle); }
    }
}
}

4.2.3 GEMM Kernel Variant — cuBLAS Backend

The existing GemmForwardKernel, GemmBackwardAKernel, GemmBackwardBKernel in trueno-gpu get a new variant that dispatches to cuBLAS instead of launching PTX. The selection is compile-time (feature flag cublas) or runtime (environment variable TRUENO_GEMM_BACKEND=cublas|ptx).

#![allow(unused)]
fn main() {
pub enum GemmBackend {
    Ptx,     // Existing hand-written PTX (fallback, reference implementation)
    Cublas,  // cuBLAS tensor core path (default when available)
}
}

4.3 Weight Storage Format Change

cuBLAS tensor core GEMMs require FP16 inputs for maximum throughput. Currently all weights are stored as FP32 on GPU. The integration requires:

1. Weight upload: Cast FP32 CPU weights to FP16 during H2D transfer
2. Gradient download: Keep FP32 for gradient accumulation and optimizer
3. Master weights: FP32 copy on CPU (already exists — CPU AdamW operates on FP32)
4. GPU weights: FP16 for forward/backward GEMMs

This is standard mixed-precision training (Micikevicius et al. 2018):

• Forward pass: FP16 weights x FP16 activations -> FP16 output
• Backward pass: FP16 weights x FP16 grad_output -> FP32 weight gradient
• Optimizer: FP32 master weights updated with FP32 gradients

4.4 Estimated Code Size

5. Benchmark Infrastructure (Raw C cuBLAS Ceiling)

5.1 Design: Three-Tier GEMM Benchmark

Following trueno’s established pattern — where raw NumPy/ndarray are the reference ceiling and Rust SIMD is measured against them — the cuBLAS integration uses raw C cuBLAS as the ceiling:

Tier 1 (CEILING):  Raw C cuBLAS    — bare cublasGemmEx(), no Rust, no wrapper
Tier 2 (TARGET):   Rust cuBLAS     — CublasHandle::gemm_f16() safe wrapper
Tier 3 (FLOOR):    Rust PTX        — GemmForwardKernel::tiled_unrolled()

FFI overhead = Tier 2 / Tier 1  (must be < 1.02x, i.e. < 2% overhead)
Speedup      = Tier 3 / Tier 2  (expect 10-50x for tensor core vs scalar)
Efficiency   = Tier 2 / peak    (target > 60% of 165 TFLOP/s = 99 TFLOP/s)

The raw C benchmark is the truth. If Tier 2 is slow, the problem is in the Rust wrapper. If Tier 1 is slow, the problem is in our cuBLAS configuration (math mode, workspace, leading dimensions). This separation is critical for root-cause analysis.

5.2 Raw C cuBLAS Benchmark

File: trueno-gpu/benchmarks/gemm_cublas_raw.c

A standalone C program that links directly against libcublas and measures isolated GEMM throughput with CUDA events (not wall clock). This is the ceiling — the best possible performance from cuBLAS on this hardware.

// trueno-gpu/benchmarks/gemm_cublas_raw.c
// Compile: nvcc -O3 -lcublas -lcuda -o gemm_cublas_raw gemm_cublas_raw.c
#include <cublas_v2.h>
#include <cuda_runtime.h>
#include <cuda_fp16.h>
#include <stdio.h>
#include <stdlib.h>

typedef struct {
    int m, n, k;
    const char* label;
} GemmShape;

// Albor training shapes (exact shapes from 350M forward+backward)
static const GemmShape SHAPES[] = {
    {4096, 1024, 1024, "attn_qkv"},      // Q/K/V projection (S=4096, H=1024)
    {4096, 4096, 1024, "ffn_gate_up"},    // FFN gate/up (S=4096, I=4096)
    {4096, 1024, 4096, "ffn_down"},       // FFN down projection
    {4096, 32768, 1024, "lm_head"},       // LM head (S=4096, V=32768)
    {1024, 1024, 1024, "square_1k"},      // Square matrix reference
    {4096, 4096, 4096, "square_4k"},      // Square matrix reference
};
#define NUM_SHAPES (sizeof(SHAPES) / sizeof(SHAPES[0]))

double benchmark_gemm(cublasHandle_t handle, int m, int n, int k,
                      int warmup, int iterations) {
    // Allocate FP16 device buffers
    half *d_A, *d_B, *d_C;
    cudaMalloc(&d_A, (size_t)m * k * sizeof(half));
    cudaMalloc(&d_B, (size_t)k * n * sizeof(half));
    cudaMalloc(&d_C, (size_t)m * n * sizeof(half));

    // Initialize with random data (via curand or host fill)
    // ... (omitted for brevity)

    float alpha = 1.0f, beta = 0.0f;

    // Warmup
    for (int i = 0; i < warmup; i++) {
        cublasGemmEx(handle, CUBLAS_OP_N, CUBLAS_OP_N,
                     m, n, k, &alpha,
                     d_A, CUDA_R_16F, m,
                     d_B, CUDA_R_16F, k,
                     &beta,
                     d_C, CUDA_R_16F, m,
                     CUBLAS_COMPUTE_32F,
                     CUBLAS_GEMM_DEFAULT_TENSOR_OP);
    }
    cudaDeviceSynchronize();

    // Timed iterations with CUDA events
    cudaEvent_t start, stop;
    cudaEventCreate(&start);
    cudaEventCreate(&stop);
    cudaEventRecord(start);

    for (int i = 0; i < iterations; i++) {
        cublasGemmEx(handle, CUBLAS_OP_N, CUBLAS_OP_N,
                     m, n, k, &alpha,
                     d_A, CUDA_R_16F, m,
                     d_B, CUDA_R_16F, k,
                     &beta,
                     d_C, CUDA_R_16F, m,
                     CUBLAS_COMPUTE_32F,
                     CUBLAS_GEMM_DEFAULT_TENSOR_OP);
    }

    cudaEventRecord(stop);
    cudaEventSynchronize(stop);
    float elapsed_ms;
    cudaEventElapsedTime(&elapsed_ms, start, stop);

    double elapsed_s = elapsed_ms / 1000.0;
    double flops = 2.0 * m * n * k * (double)iterations;
    double tflops = flops / elapsed_s / 1e12;

    cudaFree(d_A); cudaFree(d_B); cudaFree(d_C);
    cudaEventDestroy(start); cudaEventDestroy(stop);
    return tflops;
}

int main() {
    cublasHandle_t handle;
    cublasCreate(&handle);
    cublasSetMathMode(handle, CUBLAS_TENSOR_OP_MATH);

    printf("shape,m,n,k,tflops,pct_peak\n");
    for (int i = 0; i < NUM_SHAPES; i++) {
        GemmShape s = SHAPES[i];
        double tflops = benchmark_gemm(handle, s.m, s.n, s.k, 50, 1000);
        printf("%s,%d,%d,%d,%.2f,%.1f%%\n",
               s.label, s.m, s.n, s.k, tflops, tflops / 165.0 * 100.0);
    }

    cublasDestroy(handle);
    return 0;
}

Build and run:

cd trueno-gpu/benchmarks
nvcc -O3 -lcublas -lcuda -o gemm_cublas_raw gemm_cublas_raw.c
./gemm_cublas_raw > raw_cublas_baseline.csv

Expected output (RTX 4090):

shape,m,n,k,tflops,pct_peak
attn_qkv,4096,1024,1024,128.50,77.9%
ffn_gate_up,4096,4096,1024,142.30,86.2%
ffn_down,4096,1024,4096,139.80,84.7%
lm_head,4096,32768,1024,148.20,89.8%
square_1k,1024,1024,1024,85.40,51.8%
square_4k,4096,4096,4096,152.60,92.5%

This CSV becomes the performance ceiling that the Rust wrapper is measured against. If gemm_f16() is more than 2% slower than raw C, the FFI path has unnecessary overhead.

5.3 Criterion Benchmark (Rust: cuBLAS vs PTX)

File: trueno-gpu/benches/gemm_comparison.rs

Follows the exact pattern from trueno/benches/gpu_ops/matrix_benches.rs — Criterion groups with multiple backends in the same benchmark group:

#![allow(unused)]
fn main() {
// trueno-gpu/benches/gemm_comparison.rs
use criterion::{
    criterion_group, criterion_main,
    BenchmarkId, Criterion, Throughput,
};

/// Albor training shapes — exact dimensions from 350M forward/backward
const SHAPES: &[(usize, usize, usize, &str)] = &[
    (4096, 1024, 1024, "attn_qkv"),
    (4096, 4096, 1024, "ffn_gate_up"),
    (4096, 1024, 4096, "ffn_down"),
    (4096, 32768, 1024, "lm_head"),
    (1024, 1024, 1024, "square_1k"),
    (4096, 4096, 4096, "square_4k"),
];

fn bench_gemm_backends(c: &mut Criterion) {
    let mut group = c.benchmark_group("gemm");

    for &(m, n, k, label) in SHAPES {
        let flops = (2 * m * n * k) as u64;
        group.throughput(Throughput::Elements(flops));

        // Tier 2: Rust cuBLAS wrapper
        group.bench_with_input(
            BenchmarkId::new("cuBLAS", label),
            &(m, n, k),
            |bencher, &(m, n, k)| {
                let ctx = CudaContext::new(0).unwrap();
                let stream = CudaStream::new(&ctx).unwrap();
                let handle = CublasHandle::new().unwrap();
                handle.set_stream(&stream).unwrap();
                let a = GpuBuffer::random_f16(&ctx, m * k);
                let b = GpuBuffer::random_f16(&ctx, k * n);
                let mut c_buf = GpuBuffer::zeros_f16(&ctx, m * n);

                bencher.iter(|| {
                    handle.gemm_f16(m, n, k, 1.0, &a, &b, 0.0, &mut c_buf)
                        .unwrap();
                    stream.synchronize().unwrap();
                });
            },
        );

        // Tier 3: Rust PTX hand-written kernel
        group.bench_with_input(
            BenchmarkId::new("PTX", label),
            &(m, n, k),
            |bencher, &(m, n, k)| {
                let ctx = CudaContext::new(0).unwrap();
                let stream = CudaStream::new(&ctx).unwrap();
                let a = GpuBuffer::random_f32(&ctx, m * k);
                let b = GpuBuffer::random_f32(&ctx, k * n);
                let mut c_buf = GpuBuffer::zeros_f32(&ctx, m * n);
                let kernel = GemmForwardKernel::tiled_unrolled(m, n, k, 16);

                bencher.iter(|| {
                    kernel.launch(&stream, &a, &b, &mut c_buf).unwrap();
                    stream.synchronize().unwrap();
                });
            },
        );
    }

    group.finish();
}

criterion_group!(benches, bench_gemm_backends);
criterion_main!(benches);
}

Cargo.toml:

[[bench]]
name = "gemm_comparison"
path = "benches/gemm_comparison.rs"
harness = false
required-features = ["gpu", "cublas"]

Run:

cd ~/src/trueno && cargo bench --bench gemm_comparison --features "gpu,cublas"

5.4 Cross-Framework Comparison Script

File: trueno-gpu/benchmarks/gemm_comparison.py

Follows trueno/benchmarks/matmul_comparison.py — runs the raw C baseline via subprocess, parses Criterion JSON for the Rust results, and produces a unified comparison report with speedup ratios.

#!/usr/bin/env python3
"""
GEMM comparison: Raw C cuBLAS (ceiling) vs Rust cuBLAS vs Rust PTX (floor).
Follows trueno/benchmarks/matmul_comparison.py pattern.
"""
import json
import subprocess
import statistics
from pathlib import Path

SHAPES = [
    ("attn_qkv",    4096, 1024, 1024),
    ("ffn_gate_up", 4096, 4096, 1024),
    ("ffn_down",    4096, 1024, 4096),
    ("lm_head",     4096, 32768, 1024),
    ("square_1k",   1024, 1024, 1024),
    ("square_4k",   4096, 4096, 4096),
]

def run_raw_c_baseline():
    """Tier 1: Raw C cuBLAS (the ceiling)."""
    result = subprocess.run(
        ["./gemm_cublas_raw"],
        capture_output=True, text=True,
        cwd=Path(__file__).parent, timeout=300,
    )
    baselines = {}
    for line in result.stdout.strip().split("\n")[1:]:  # Skip CSV header
        parts = line.split(",")
        label, tflops = parts[0], float(parts[4])
        baselines[label] = tflops
    return baselines

def load_criterion_results():
    """Tier 2 + 3: Parse Criterion JSON from target/criterion/."""
    criterion_dir = Path("target/criterion/gemm")
    results = {"cuBLAS": {}, "PTX": {}}
    for estimates in criterion_dir.rglob("estimates.json"):
        with open(estimates) as f:
            data = json.load(f)
        mean_ns = data["mean"]["point_estimate"]
        # Extract backend and shape from path
        parts = estimates.parts
        backend = parts[-4]   # "cuBLAS" or "PTX"
        shape = parts[-3]     # "attn_qkv", etc.
        results[backend][shape] = mean_ns
    return results

def compute_tflops(shape_label, time_ns):
    """Convert mean time to TFLOP/s."""
    for label, m, n, k in SHAPES:
        if label == shape_label:
            flops = 2.0 * m * n * k
            return flops / (time_ns * 1e-9) / 1e12
    return 0.0

def format_ratio(ratio):
    if ratio < 1.02:
        return f"  {ratio:.3f}x (within 2%)"
    elif ratio < 1.10:
        return f"  {ratio:.3f}x (within 10%)"
    else:
        return f"  {ratio:.3f}x SLOW"

def main():
    raw_c = run_raw_c_baseline()
    criterion = load_criterion_results()

    print("=" * 78)
    print("GEMM BENCHMARK: Raw C cuBLAS (ceiling) vs Rust cuBLAS vs PTX (floor)")
    print("=" * 78)
    print()
    print(f"{'Shape':<14} {'Raw C':>10} {'Rust cuBLAS':>12} {'PTX':>10} "
          f"{'FFI OH':>8} {'Speedup':>8} {'% Peak':>8}")
    print("-" * 78)

    for label, m, n, k in SHAPES:
        raw_tflops = raw_c.get(label, 0)

        cublas_ns = criterion["cuBLAS"].get(label)
        cublas_tflops = compute_tflops(label, cublas_ns) if cublas_ns else 0

        ptx_ns = criterion["PTX"].get(label)
        ptx_tflops = compute_tflops(label, ptx_ns) if ptx_ns else 0

        ffi_overhead = cublas_tflops / raw_tflops if raw_tflops > 0 else 0
        speedup = cublas_tflops / ptx_tflops if ptx_tflops > 0 else 0
        pct_peak = cublas_tflops / 165.0 * 100

        print(f"{label:<14} {raw_tflops:>8.1f}T  {cublas_tflops:>10.1f}T  "
              f"{ptx_tflops:>8.1f}T  {1/ffi_overhead:>7.3f}x {speedup:>7.1f}x "
              f"{pct_peak:>6.1f}%")

    print()
    print("FFI OH = Raw C / Rust cuBLAS (< 1.02x = good)")
    print("Speedup = Rust cuBLAS / PTX")
    print("% Peak = Rust cuBLAS / 165 TFLOP/s (RTX 4090 FP16)")

if __name__ == "__main__":
    main()

Expected report:

==============================================================================
GEMM BENCHMARK: Raw C cuBLAS (ceiling) vs Rust cuBLAS vs PTX (floor)
==============================================================================

Shape          Raw C   Rust cuBLAS       PTX   FFI OH  Speedup   % Peak
------------------------------------------------------------------------------
attn_qkv       128.5T       127.8T      2.1T   1.005x    60.9x    77.5%
ffn_gate_up    142.3T       141.5T      2.3T   1.006x    61.5x    85.8%
ffn_down       139.8T       138.9T      2.2T   1.006x    63.1x    84.2%
lm_head        148.2T       147.1T      1.9T   1.007x    77.4x    89.2%
square_1k       85.4T        84.8T      1.5T   1.007x    56.5x    51.4%
square_4k      152.6T       151.8T      2.5T   1.005x    60.7x    92.0%

FFI OH = Raw C / Rust cuBLAS (< 1.02x = good)
Speedup = Rust cuBLAS / PTX
% Peak = Rust cuBLAS / 165 TFLOP/s (RTX 4090 FP16)

5.5 Regression Detection

File: trueno-gpu/benchmarks/check_gemm_regression.py

Follows trueno/scripts/check_regression.py — saves baselines with git metadata, compares current runs, and fails CI on regressions.

Thresholds (adapted for GPU benchmarks which have higher variance):

Baseline capture:

# Save baseline with hardware metadata
cd trueno-gpu
./benchmarks/save_gemm_baseline.sh
# Saves to .performance-baselines/gemm-baseline-current.csv
# Header: commit, branch, date, GPU (nvidia-smi), CUDA version, driver version

Regression check:

# Compare current run against baseline
./benchmarks/check_gemm_regression.py \
    --baseline .performance-baselines/gemm-baseline-current.csv \
    --current /tmp/gemm-bench-current.csv \
    --regression-threshold 0.10 \
    --warning-threshold 0.05

5.6 Makefile Targets

Following trueno’s Makefile convention:

# trueno-gpu/Makefile (new targets)

bench-gemm:                  ## Full GEMM benchmark (cuBLAS vs PTX)
	cargo bench --bench gemm_comparison --features "gpu,cublas"

bench-gemm-raw:              ## Raw C cuBLAS ceiling benchmark
	cd benchmarks && nvcc -O3 -lcublas -lcuda -o gemm_cublas_raw gemm_cublas_raw.c
	cd benchmarks && ./gemm_cublas_raw

bench-gemm-compare:          ## Three-tier comparison report
	$(MAKE) bench-gemm-raw
	$(MAKE) bench-gemm
	cd benchmarks && python3 gemm_comparison.py

bench-gemm-baseline:         ## Save current results as baseline
	$(MAKE) bench-gemm-compare
	./benchmarks/save_gemm_baseline.sh

bench-gemm-regression:       ## Check for regressions against baseline
	$(MAKE) bench-gemm-compare
	./benchmarks/check_gemm_regression.py \
		--baseline .performance-baselines/gemm-baseline-current.csv \
		--current /tmp/gemm-bench-current.csv

5.7 Contract Integration

The benchmark infrastructure maps directly to contract obligations:

Add to cublas-gemm-v1.yaml:

  ffi_overhead:
    formula: |
      overhead = T_rust_cublas / T_raw_c_cublas
      For identical GEMM shape, same GPU, same cuBLAS config.
    invariants:
      - "overhead < 1.02 for all training shapes (< 2% FFI tax)"
      - "Measured via CUDA events, not wall clock"
      - "Warmup: 50 iterations discarded before measurement"

# Additional falsification test:
  - id: FALSIFY-CUBLAS-008
    rule: "Rust cuBLAS FFI overhead < 2%"
    prediction: "T_rust / T_raw_c < 1.02 for all 6 training shapes"
    test: |
      Run gemm_cublas_raw (C) and gemm_comparison (Criterion) on same GPU.
      Compare TFLOP/s for each shape. Ratio must be > 0.98.
    if_fails: "Unnecessary copies, redundant stream syncs, or Rust allocation overhead in wrapper"

6. Implementation Phases (Contract-Driven)

Every phase follows the same discipline:

pv validate   -> implement -> probador verify -> renacer trace -> pv audit
                              bench-gemm-compare (three-tier)

Phase 0: Baseline Measurement

Contract: training-step-budget-v1.yaml Tool: renacer BrickTracer + probador brick budgets + raw C cuBLAS ceiling

1. Run raw C cuBLAS benchmark to establish the hardware ceiling per shape
2. Instrument train_step_single() with BrickTracer spans for every component
3. Run 50-step profiling on 350M with PTX backend
4. Confirm step time breakdown matches estimates in section 2.4
5. Establish brick budgets as probador assertions
6. Save baselines: make bench-gemm-baseline
7. This becomes the floor + ceiling that all phases are measured against

Renacer layer tracing output (per-block detail):

albor-baseline / training-step [4400ms]
+-- embed_forward [180ms]
+-- pcie_h2d_hidden [12ms]
+-- block_0_forward [95ms]
|   +-- gemm_qkv [42ms]         # 3 GEMMs: Q, K, V projections
|   +-- attention_scores [8ms]   # QK^T GEMM
|   +-- attention_output [14ms]  # attn_out GEMM
|   +-- ffn_forward [28ms]       # 3 GEMMs: gate, up, down
|   +-- rmsnorm [3ms]
+-- block_0_backward [190ms]
|   +-- gemm_backward [165ms]    # 14 weight + activation GEMMs
|   +-- elementwise [25ms]       # SiLU backward, RMSNorm backward
+-- block_0_optimizer [33ms]     # CPU AdamW (D2H + update + H2D)
+-- ... (blocks 1-23)
+-- lm_head_forward [45ms]
+-- pcie_d2h_logits [35ms]
+-- cross_entropy [22ms]
+-- pcie_h2d_grad_logits [35ms]
+-- lm_head_backward [90ms]

Each span is an OTLP trace viewable in Jaeger. Anomalous spans (CV > 15%) trigger automatic escalation to syscall-level profiling.

Phase 1: FFI + Forward Pass — COMPLETE

Contract: cublas-gemm-v1.yaml (FALSIFY-CUBLAS-001, -003, -008) Status: ✅ Implemented in trueno#165, entrenar#231

1. ✅ cublas_sys.rs: FFI bindings (libloading + OnceLock, ~270 lines)
2. ✅ cublas.rs: Safe RAII wrapper with gemm_f32(), gemm_f16(), row-major helpers
3. ✅ Forward GEMM dispatch: cuBLAS when available, PTX fallback transparent
4. ✅ Verified: 152.3 TFLOP/s isolated (FALSIFY-CUBLAS-003), loss matches PTX

Phase 2: Backward Pass — COMPLETE

Contract: cublas-gemm-v1.yaml (FALSIFY-CUBLAS-002, -006, -007) Status: ✅ Implemented in entrenar#231

1. ✅ cublas_gemm_backward_a(): Trans/NoTrans cuBLAS dispatch
2. ✅ cublas_gemm_backward_b(): NoTrans/Trans cuBLAS dispatch
3. ✅ Gradient accumulation stays FP32 (cuBLAS uses FP32 compute)
4. ✅ Verified: 50M 5-step regression — loss 10.41 (was 10.39), all params get gradients

Phase 3: Optimization — COMPLETE

Contract: training-step-budget-v1.yaml (FALSIFY-BUDGET-001, -002) Status: ✅ Verified on 50M and 350M

1. ✅ CUBLAS_TENSOR_OP_MATH enabled (TF32 tensor cores on sm_89)
2. ✅ cuBLAS handle reused across steps (RAII, one per cache)
3. ✅ Stream binding once per step (set_forward_cublas_stream)
4. ✅ Measured results:
   • 50M: 1,744 tok/s (was 890), 293ms/step (was 575ms), 1.96x
   • 350M: 1,485 tok/s (was 934), 1,379ms/step (was 4,400ms), 3.19x
   • VRAM: +4 MB overhead (negligible)

6. Performance After cuBLAS (Measured)

6.1 Measured Throughput (Phase 1-3 Complete)

cuBLAS integration verified on both 50M and 350M models (RTX 4090, seq=1024, batch=4):

50M model (12 layers, hidden=512):

350M model (24 layers, hidden=1024, seq=512, batch=4):

Verified via apr train apply --config pretrain-350m-cuda-test.yaml (entrenar PR #233).

350M step budget (cuBLAS):
  GEMM compute:     ~500 ms (was ~2500 ms with PTX — 5x speedup on large matrices)
  Attention (PTX):  ~400 ms (batched_4d_gemm, still scalar)
  CPU optimizer:    ~300 ms (D2H + AdamW + H2D per block)
  Elementwise:      ~100 ms (RMSNorm, SiLU, residual, etc.)
  PCIe transfers:   ~136 ms (embed H2D + grad transfers)
  Total:            ~1436 ms/step

Note: Attention GEMMs (batched_4d_gemm_forward) remain PTX. Converting these to cublasGemmStridedBatched would give an additional 1.3-1.5x.

6.2 cuBLAS Raw Capability

Measured with bench_cublas_vs_ptx example (isolated, no training overhead, TF32 mode):

Key findings:

• 12-27x kernel-level speedup (cuBLAS TF32 vs scalar PTX FP32)
• Large training shapes (>1024) achieve 80-86% of TF32 tensor core peak (165 TFLOP/s)
• GQA thin-matrix shape [4096,256,1024] achieves only 45% peak (memory-bandwidth bound)
• End-to-end training speedup is 3.06x because GEMMs are only part of the step

6.3 MFU Analysis (Post-cuBLAS, Measured)

50M model (measured):
  FLOPs per step:     6 × 62M × 4096 = 1.52 TFLOP
  Step time:          293 ms
  Achieved FLOP/s:    1.52 / 0.293 = 5.19 TFLOP/s
  MFU (vs FP16):      5.19 / 165 = 3.1%
  MFU (vs FP32):      5.19 / 82.6 = 6.3%

350M model (measured, seq=512, batch=4):
  FLOPs per step:     6 × 370M × 2048 = 4.55 TFLOP
  Step time:          1,379 ms (measured, not projected)
  Achieved FLOP/s:    4.55 / 1.379 = 3.30 TFLOP/s
  MFU (vs FP16):      3.30 / 165 = 2.0% → reported as 4.3% (runtime measurement includes seq_len scaling)
  MFU (vs FP32):      3.30 / 82.6 = 4.0%

After cuBLAS fixes the linear GEMM bottleneck, the attention GEMMs (PTX) and CPU optimizer become the dominant bottlenecks (~400ms + ~300ms = ~700ms of 1379ms). To reach research-grade MFU, further phases are needed:

6.4 Full Optimization Path

*Phase 5a: 257ms uses seq=512 profile config vs seq=1024 for Phases 1-4. TF32 provides 0% measurable improvement at 350M (compute <15% of step time).

*Phase 5b measured at seq=1024 (production config). Step 1 = 444ms (async) / 638ms (blocking, true GPU time). Includes JIT warmup (~200ms). Forward GPU time 347ms → 14ms (24.8x) at seq=512. At seq=1024: 9,216 tok/s (9.9x vs baseline). 100,352 kernel launches → ~550 (182x fewer). nsys-verified.

Fused QKV (originally Phase 5): CANCELLED — all GEMMs already use cuBLAS. Identical FLOP count, negligible dispatch saving (0.1%), high implementation cost.

Current position: Phase 5b achieves 26.7% MFU at seq=1024 — within 2x of research-grade throughput. Remaining bottleneck is per-kernel dispatch overhead (~550 launches/step) and host↔device synchronization.

Each future phase gets its own contract before implementation begins.

6.5 Phase 4 Results: Attention GEMMs (MEASURED)

cuBLAS cublasSgemmStridedBatched replaces hand-written PTX for multi-head attention score computation (QK^T and attn·V). Implemented in trueno-gpu 0.4.25

• entrenar PR #234 (merged).

Measured results (350M, seq=512, batch=4, RTX 4090):

Analysis: The improvement is modest (2.3%) because at seq=512 the attention matrices are small (512×512×64 per head, batch_count=64). At seq=1024 or seq=2048 the improvement would be larger as attention GEMMs scale as O(seq²).

Implementation (trueno-gpu 0.4.25, entrenar PR #234):

• cublasSgemmStridedBatched FFI in trueno-gpu cublas_sys.rs
• Safe wrapper gemm_f32_strided_batched_row_major() in cublas.rs
• batch_count = batch_size * num_heads (4 × 16 = 64)
• Fast path in batched_4d_gemm_forward with PTX fallback

6.6 Step Time Profiling (KAIZEN-047, MEASURED)

Per-phase wall-clock breakdown from StepProfiler (KAIZEN-047). Profiled on 350M model, seq=512, batch=4, RTX 4090, cuBLAS enabled. Combined forward-only (NaN-skipped) and full forward+backward samples.

Forward-only steps (200 profiled samples, avg 255.7 ms/step):

Full forward+backward step (1 sample, 323 ms):

Key finding: Forward pass dominates at 80-94% of step time. Each block dispatches ~20 GPU operations (7 GEMMs + attention pipeline + norms + activations

• residual adds) = 480+ kernel launches per step.

Critical observation: ALL GEMMs already use cuBLAS (Phase 1-4, ALB-075): forward gemm_forward, backward gemm_backward_a/gemm_backward_b, AND attention batched cublasSgemmStridedBatched. There are no remaining PTX GEMMs in the training loop.

Anomaly: The forward phase measures 240ms of CPU wall-clock time for what should be purely async GPU dispatches. At ~5μs per cuBLAS dispatch for ~480 operations, expected CPU time is ~2.4ms — a 100x discrepancy. Possible causes:

1. CUDA command queue backpressure (driver blocks CPU when queue is full)
2. Implicit cuBLAS synchronization between GEMMs on the same stream
3. cuBLAS workspace allocation/reallocation between differently-sized GEMMs
4. Kernel cache mutex contention (unlikely — single-threaded)

Fused QKV analysis (CANCELLED): Since all GEMMs use cuBLAS, merging 3 QKV GEMMs into 1 fused GEMM yields identical FLOP count and saves only 2 dispatches per block (48 total, ~240μs, 0.1% of step time). The implementation requires GPU split/concat kernels, backward pass rewrite, and optimizer restructuring. Cost-benefit ratio is unfavorable.

Next bottleneck: Not dispatch count, not CPU optimizer — it’s understanding why async GPU dispatches appear to block the CPU for 240ms. Requires nsys profiling or CUDA_LAUNCH_BLOCKING=1 timing.

Optimization targets (revised):

1. nsys profiling — identify actual GPU kernel vs idle vs sync time
2. Reduce implicit synchronization — eliminate any cuBLAS sync barriers
3. CUDA Graphs — capture forward/backward as graph, eliminate per-kernel dispatch
4. Kernel fusion — merge element-wise ops (residual_add + RMSNorm) to reduce memory traffic

6.7 Fused QKV Analysis (CANCELLED)

Phase 5 was originally planned as fused QKV projection (3 GEMMs → 1 per block). Analysis during implementation revealed this is not impactful:

Why fused QKV doesn’t help:

1. All GEMMs already use cuBLAS (ALB-075, Phases 1-4). Forward, backward, and attention batched GEMMs all dispatch via tensor core paths.
2. Identical FLOP count: 3 separate GEMMs (Q, K, V) = 1 fused GEMM in total floating point operations. No compute savings.
3. Negligible dispatch saving: 48 fewer kernel launches × ~5μs = 240μs. Against a 240ms forward pass, this is 0.1% improvement.
4. High implementation cost: Requires GPU split/concat kernels (trueno lacks cuMemcpy2D), backward pass rewrite (concatenated gradient assembly), optimizer restructuring (merged w_qkv states), and checkpoint format changes.
5. GQA complicates layout: Q dim (1024) ≠ K/V dim (256), so the output [seq, 1536] cannot be trivially sliced without strided copies.

What matters instead: The 240ms forward measurement is 100x slower than expected for async GPU dispatches. Understanding and fixing this anomaly would yield far greater improvement than any kernel-level fusion.

6.8 Forward Pass Anomaly — ROOT CAUSE FOUND (ALB-076, FIXED)

Observation: The StepProfiler measures 240ms of CPU wall-clock time for the 24-block forward loop. Expected CPU dispatch time: ~2.4ms. nsys profiling was used to identify the root cause.

nsys profiling results (50 steps, RTX 4090):

GPU Kernel Time Breakdown (nsys --stats=true):
  97.1%  46.6s  5,017,600 instances  rmsnorm          avg=9.3μs
   0.8%   0.4s      9,600 instances  cutlass GEMM     avg=37.8μs
   0.6%   0.3s     19,200 instances  cutlass GEMM     avg=14.1μs
   0.4%   0.2s      4,800 instances  cutlass GEMM     avg=42.3μs
   ...remaining kernels < 0.2% each

Root cause: Per-row RMSNorm kernel launches

The rms_norm_forward() in normalization.rs launched RmsNormKernel in a CPU loop:

#![allow(unused)]
fn main() {
// BEFORE (97.1% of GPU time):
let config = LaunchConfig { grid: (1, 1, 1), block: (32, 1, 1), shared_mem: 0 };
for batch_idx in 0..batch_size {  // 2,048 iterations per norm call!
    stream.launch_kernel(module, kernel_name, &config, &mut args)?;
}
}

• 49 norm calls/step × 2,048 launches each = 100,352 kernel launches/step
• Each launch: grid=(1,1,1), block=(32,1,1) = 1 warp on 1 SM out of 128
• At ~9.3μs per launch: 933ms of GPU time per step just in RMSNorm
• Meanwhile, all cuBLAS GEMMs total only ~22ms per step

Five Whys:

1. Why is forward 240ms? GPU backpressure from 100K RMSNorm kernel launches
2. Why 100K launches? rms_norm_forward loops batch_size=2048 times
3. Why per-row loop? RmsNormKernel processes one row (grid=(1,1,1))
4. Why single-row kernel? Written before BatchedVectorizedRmsNormKernel
5. Why not updated? Backward module already used batched variant; forward wasn’t

Fix (entrenar PR #238, merged):

#![allow(unused)]
fn main() {
// AFTER (single launch, all rows in parallel):
let kernel = BatchedVectorizedRmsNormKernel::new(hidden_size, batch_size);
let config = LaunchConfig {
    grid: (1, batch_size, 1),  // One block per row
    block: (256, 1, 1),        // 8 warps per block
    shared_mem: 8 * 4,
};
stream.launch_kernel(module, "batched_rmsnorm_vectorized", &config, &mut args)?;
}

Measured impact (350M, seq=512, batch=4, RTX 4090):