428 Commits

This Branch

This Branch

All Branches

Author	SHA1	Message	Date
osobhandClaude Fable 5	ad6405663f	fix(streaming): sane DynamicBatchingConfig default; worker lifecycle regression tests ... The new lifecycle tests caught that the batch-optimizer worker crashed at spawn: DynamicBatchingConfig derived Default (all zeros), and tokio's interval() panics on a zero period. Default is now a usable config (batch 32 in [1,128], step 4, 1ms latency target, 100-sample window, 1s optimization interval, AIMD adaptation). New regression tests in AdaptiveProcessor, EdgeComputingManager, and MonitoringSystem assert that all workers are still alive shortly after start() (catches workers dying at startup) and that stop() completes via the graceful control-channel path, not the 5s abort backstop (catches shutdown hangs). cargo test -p rtx-streaming: 58 lib + 8 integration + 6 aux, all green. Co-Authored-By: Claude Fable 5 <[email protected]>	2026-07-10 17:16:04 -07:00
osobhandClaude Fable 5	c83e0fb22d	fix(streaming): wire worker control planes for real graceful shutdown ... Follow-up to a0bf294, which tolerated dead control channels; this makes them functional: - AdaptiveProcessor: mpsc control channel (single consumer behind a mutex, broke on ANY message including Start) replaced with broadcast; all three workers (resource monitor, batch optimizer, pressure monitor) subscribe and exit only on ControlCommand::Stop - EdgeComputingManager / MonitoringSystem: their 7 interval-loop workers now subscribe to the existing broadcast control channels and exit on Stop instead of looping forever - stop() in all three: graceful join with 5s timeout, abort only as a backstop (previously unconditional abort mid-tick) - benches: criterion needs async_tokio for Bencher::to_async — bench target now compiles (clippy --all-targets clean) cargo test -p rtx-streaming: 55 lib + 8 integration + 6 aux, all green. Co-Authored-By: Claude Fable 5 <[email protected]>	2026-07-10 17:04:35 -07:00
osobhandClaude Fable 5	a0bf29461b	fix(streaming): real inference backend wiring and lifecycle fixes; full suite green ... - token_generator: backend is now an optional real rtx-inference engine (RwLock<Option<Arc<InferenceEngine>>>) with ServingTokenizer support; set_backend/set_tokenizer plumbing through StreamingServer - connection_manager: ConnectionPool::acquire no longer errors when the idle cache is full — creates fresh connections up to max_connections - streaming_server: ServerState::Running on construction; stream_inference generates one token per step (chunk_size semantics) - lifecycle bugs surfaced by the newly-compiling integration tests: * start(): broadcast control-channel send with zero subscribers was treated as fatal ("channel closed") in RealtimePipeline, EdgeComputingManager, MonitoringSystem — now tolerated * stop(): AdaptiveProcessor/EdgeComputingManager/MonitoringSystem awaited worker interval loops that never exit (test hung 5h) — workers are now aborted with cancellation-aware join - integration_tests: removed stale .await on now-synchronous methods cargo test -p rtx-streaming: 55 lib + 8 integration + 6 aux, all passing. Co-Authored-By: Claude Fable 5 <[email protected]>	2026-07-10 15:48:33 -07:00
osobhandClaude Fable 5	73102b71cf	feat(transformers): real compute in orchestrator modalities and attention planner ... - orchestrator_core: execute_classical is genuine seeded QKV self-attention; execute_flash_attention delegates to rtx_flash_attention::flash_attention_forward (8-head reshape, clear Err on indivisible hidden); execute_hybrid composes the two; execute_edge/execute_distributed return explicit "modality not implemented" errors instead of fake success — which makes the previously-dead fallback_modalities retry loop real. - tensor_core_kernels: AttentionComputationOptimizer::execute dispatches Standard/Online to scaled_dot_product_attention, Flash to rtx-flash-attention, Approximated to an explicit Err (no approximation kernel exists; refuses to compute exact attention under an approximated label). - 10 new always-on tests incl. flash-vs-standard 1e-3 agreement and an Edge->Distributed->Classical fallback end-to-end. - docs/consolidation.md updated: both entries moved from scaffolding/no-op-stub status to real dispatch descriptions. 982 rtx-transformers lib tests pass. Co-Authored-By: Claude Fable 5 <[email protected]>	2026-07-10 07:06:50 -07:00
osobhandClaude Fable 5	fce6cef262	docs: JEPA roadmap — GPU resume re-upload done; remaining items need multi-GPU ... Co-Authored-By: Claude Fable 5 <[email protected]>	2026-07-10 03:59:48 -07:00
osobhandClaude Fable 5	a755627269	feat(jepa): GPU weight re-upload on checkpoint resume ... GpuViTEncoder gains upload_weights (extracted from construction), cpu_weights_mut, and reupload_weights; JepaTrainerV2 exposes context_encoder_as_any_mut for backend-specific downcasts. The runner resume path now restores checkpoint fields into the GPU encoder's host copy and pushes them back to the device buffers — previously GPU resume restored only the step counter with a warning. If re-upload fails after host restore, the run aborts rather than training on stale device weights. Verified live on the RTX 5060 Ti: train 20 steps -> resume from the .jepa binary with total_steps=30 -> "Resumed from step 20", exactly 10 further steps, eval runs, no warnings. New tests: GPU output changes after host mutation + re-upload; CPU-target re-upload is a no-op Ok. 972 CPU tests / 37 GPU jepa_gpu tests pass. Co-Authored-By: Claude Fable 5 <[email protected]>	2026-07-10 03:59:36 -07:00
osobhandClaude Fable 5	74d3db7ee7	docs: refresh honesty notes for rewired demos, consolidation audit, ServingTokenizer ... Co-Authored-By: Claude Fable 5 <[email protected]>	2026-07-10 03:52:13 -07:00
osobhandClaude Fable 5	68481ef314	docs: JEPA roadmap round-2 done items; queue GPU-resume re-upload and TP/PP ... Co-Authored-By: Claude Fable 5 <[email protected]>	2026-07-10 03:51:50 -07:00
osobhandClaude Fable 5	ac3f2af06b	feat(jepa): eval in the training loop, NCCL GPU AllReduce, GPU checkpointing ... - Eval: run_jepa_training now runs k-NN (k=5) + linear-probe evaluation every eval_every steps and at the end (deduped when aligned); JepaEvalResult recorded in JepaTrainingSummary (final_knn_acc / final_probe_acc), printed by the CLI, appended as an # eval section to the metrics CSV. Probe set is deterministic LCG synthetic (offset seed, never aliases training batches) or real shard labels when loaded. - NCCL: real GPU-direct AllReduce backend behind the new `nccl` feature (cudarc/nccl, dlopen-based so builds don't need libnccl). NCCL unique id is bootstrapped over the existing TCP rendezvous (master_port+137); data path is htod -> ncclAllReduce(Sum) -> dtoh -> mean. catch_unwind guards cudarc's panic-on-missing-lib so training falls back instead of aborting. Verified for real on the RTX 5060 Ti: single-rank GPU all_reduce identity test passes (26/26 with --features nccl). - GPU checkpointing/eval: JepaTrainerV2::context_encoder_cpu_weights() exposes host-side weights for both CPU and GPU encoders (GpuViTEncoder::cpu_weights); checkpoint save and eval now work for GPU training runs (verified: .jepa binaries written and 2 eval passes during a live GPU CLI run). Resume with a GPU encoder restores the step counter and warns that weight re-upload is not yet implemented rather than silently training on stale weights. 125 runner/distributed/vit tests pass; CLI 8/8; cuda check clean. Co-Authored-By: Claude Fable 5 <[email protected]>	2026-07-10 03:51:36 -07:00
osobhandClaude Fable 5	f6b2308381	docs: JEPA roadmap — 2026-07-10 items done, next tier queued ... Co-Authored-By: Claude Fable 5 <[email protected]>	2026-07-10 00:04:52 -07:00
osobhandClaude Fable 5	19b6581f9c	feat(jepa): extended GPU training, data pipeline, integration, and cargo config ... - jepa_gpu: remove the bring-up 2-block cap; full-depth GPU-resident ViT verified against a full-depth CPU reference on the RTX 5060 Ti (depth-12 ViT-Tiny max_rel_err <= 6.6e-5). CpuViTEncoder's own hidden min(depth,2) cap removed too — CPU-path callers now get the model they configured. - jepa_distributed: real TCP parameter-server AllReduce backend (rendezvous handshake with world-size/rank validation, length- prefixed f32 payloads, connect/read/accept timeouts, connect retry until deadline so early peers survive rank 0 still computing); jepa_runner wires it for world_size > 1 and fails hard on collective errors. Two-rank loopback training run covered by test. - rtx-jepa-cli (new crate): rtx-jepa binary with train/bench/plan/ validate subcommands driving JepaRunConfig, run_jepa_training, run_jepa_benchmark, and ClusterTrainingPlan (plan --emit-config round-trips through a config serializer). GPU bench on this node: 86k patches/sec vs 1.3k CPU (~64x). - ViTSizeStr::Micro (d=32, depth=2) added as an explicit test/smoke size now that no hidden caps keep full-size configs cheap; heavy tests moved onto it (rtx-transformers suite: 367s -> 5s, and the runner subset had ballooned to 35min at full depth before this). 966 lib tests pass; 35/35 jepa_gpu with cuda; 8/8 CLI tests. Co-Authored-By: Claude Fable 5 <[email protected]>	2026-07-10 00:04:21 -07:00
osobhandClaude Fable 5	b6440905e7	feat(jepa): gzip-compressed WebDataset shard support ... read_webdataset_shard detects the gzip magic bytes (1F 8B, not extension) and decompresses via flate2 before tar parsing; WebDatasetShard::load no longer rejects .tar.gz/.tgz. Round-trip test writes a real gzipped tar and loads it back. Co-Authored-By: Claude Fable 5 <[email protected]>	2026-07-09 22:37:05 -07:00
osobhandClaude Fable 5	4ba1b78215	docs(consolidation): audit outcome — flagged MoE/flash-attn duplicates are not duplicates ... Call-site rewrite pass audited every flagged site; none needed consolidation: metal_moe is an API-consistent backend specialization, modular/router.rs is module-level (not expert-token) routing, glam.rs is disabled dead code with a pre-existing bug (noted for whoever re-enables it), and the three flash-attention "reimplementations" turn out to be planner scaffolding, no-op stubs, and a doc comment — no attention math exists to delegate. jepa_gpu's attention is documented as part of the fused GPU ViT block by design. Verified no regressions: rtx-transformers 961 lib tests pass, jepa_gpu 34/34 with cuda, rtx-training cuda check clean. Co-Authored-By: Claude Fable 5 <[email protected]>	2026-07-09 22:30:35 -07:00
osobhandClaude Fable 5	fdc1432072	feat(jepa): full GPU-resident ViT block on CUDA ... jepa_gpu.rs rewrite: all GEMMs (QKV, attention scores via a single cuBLAS call with folded 1/sqrt(dk) scale, weighted sum, projections, FFN) plus layernorm/GELU/bias/residual kernels now operate on device-resident CudaSlice buffers; new nvrtc kernels for numerically stable row softmax and head extract/scatter replace the CPU reorder loops. Two host transfers remain per encode(): patch-token upload and final output download. Fallback is now genuine-unavailability only (no device / feature off / kernel compile failure); per-op errors in live GPU mode are hard errors instead of silent per-op CPU downgrades. Parity vs CpuViTEncoder verified on RTX 5060 Ti / CUDA 13.1: max_rel_err <= 3.1e-5 across tiny/Tiny-192 configs (tolerance 1e-3). 34/34 jepa_gpu tests pass with --features cuda; 33/33 CPU-only. CLAUDE.md JEPA "Next" list updated to reflect completed GPU wiring, WebDataset reading, and cluster-plan consumption. Co-Authored-By: Claude Fable 5 <[email protected]>	2026-07-09 22:10:27 -07:00
osobhandClaude Fable 5	e080748d88	feat(demos,inference): wire simulation demos to real compute; fix embedding lookup and weight-name aliases ... Demos: - rtx-distllm-demo: real rtx-tensor weights per shard, real scaled-dot-product attention forward, metrics measured (Instant) instead of hardcoded constants; network topology remains a documented simulation fed by real tensor byte sizes. - rtx-model-zoo: MockInferenceEngine deleted; RealInferenceEngine loads a tiny real transformer into rtx_inference::InferenceEngine and runs genuine engine.infer per request; domain outputs are explicitly- labeled toy proxies derived from real output tokens. - rtx-inference-profiler: mock models deleted; profiles real matmul/softmax pipelines on rtx-tensor with measured latency/memory. Inference-path bugs the demos surfaced (fixed here): - ForwardPass::apply_embedding misused Tensor::gather for the embedding lookup — gather returns the indices' shape, silently dropping the hidden dim and breaking every downstream broadcast. Now uses the existing Tensor::embedding_lookup ([vocab,hidden] x [batch,seq] -> [batch,seq,hidden]). - Attention weight lookup accepts both self_attn. (HF-LLaMA) and attention. prefixes; final layer norm accepts norm.weight / model.norm.weight / ln_f.weight aliases. - Integration fixture gains the final norm weight; the previously always-failing engine tests now pass (8/8 model_loading_test). End-to-end inference through the real engine now works for the first time — verified via model_zoo_demo producing real forward-pass outputs across all categories. Co-Authored-By: Claude Fable 5 <[email protected]>	2026-07-09 22:06:29 -07:00
osobhandClaude Fable 5	733b02cd8b	feat(inference): concrete EAGLE draft model + real tokenizer at the serving boundary ... EAGLE (rtx-inference/src/eagle.rs, ~610 lines, mirrors medusa.rs conventions): EagleDraftHead autoregressive FFN with Concat/Add/ Attention feature fusion, EagleHeads draft model with draft/ draft_steps (per-step top-k for candidate trees) and teacher-forced training_loss; implements the speculative::EagleDraftModel trait so it plugs into the orchestration layer. 38 unit tests. Tokenizer (rtx-inference/src/tokenizer.rs): ServingTokenizer enum — Vocab (HuggingFace tokenizers, loadable from tokenizer.json) or ByteLevel fallback preserving previous behavior. rtx-serving-api's AppState and rtx-streaming's token generator now encode/decode through it (with_engine_and_tokenizer / set_tokenizer added; existing signatures unchanged). Also fixes two pre-existing compile errors in rtx-streaming (missing import, stray .await) that blocked its lib tests entirely. Tests: rtx-inference 328 pass, rtx-serving-api 193 pass, rtx-streaming 53 pass (2 pre-existing mock-server connection failures unrelated to these changes). Co-Authored-By: Claude Fable 5 <[email protected]>	2026-07-09 21:54:05 -07:00
osobhandClaude Fable 5	0cbfc1a739	fix(tests): repair rtx-onnx-codegen build and all pre-existing test failures in rtx-serving-api and rtx-runtime ... - rtx-onnx-codegen: re-export AttributeValue from ir (private-module import broke the whole crate; remaining errors were knock-ons). - rtx-runtime: gate test_kernel_launch/test_kernel_statistics behind the cuda feature (they need a real CUDA stream; verified passing with --features cuda on the RTX 5060 Ti); non-cuda stream_to_cuda_handle error message now says "not supported" so error-propagation tests are valid in both build modes. - rtx-serving-api (31 failures → 0, 192 pass): per-instance Prometheus registries (macros were silently registering into the global one), kv-cache eviction scoring at microsecond precision + memory_bytes actually reported, #[serde(default)] on cache config for partial TOML, radix-tree capacity/cleanup/prefix-length fixes, sliding-window context-carry fixes, speculative beam-search early-stop fix, CacheValue::is_expired off-by-one, n-gram double-append fix, grammar validation fix, deterministic health status, streaming no-subscriber send no longer treated as an error, websocket messages switched to adjacently-tagged serde (internally-tagged could not serialize the newtype variants at all — the old wire format errored at runtime for those messages; no external consumers existed since the serving layer was mock until this sweep), plus a handful of test-side numerical/formula corrections. Co-Authored-By: Claude Fable 5 <[email protected]>	2026-07-09 19:49:01 -07:00
osobhandClaude Fable 5	5f32165184	chore(sweep): delete 43 orphaned source files; document SYCL/demo/duplication status ... Deletions (all verified unreferenced by any mod/include/path declaration; git history preserves them): - rtx-transformers: entire orphaned curriculum/ split (mod.rs holds the real inline implementation), non-_simple graph variants, superseded simmim/jepa_integration files, layers/{sliding_window_attention, positional_encoding,ssm_state_cache_original}, lib_full/lib_minimal/ error_full/error_minimal, orphaned MoE impls (moe_layer, moe_integration). - rtx-distributed/parallel_old.rs; rtx-flash-attention/{core_full, lib_full}.rs; rtx-compress legacy_distillation + structured_pruner. - rtx-tensor/tensor_core.rs; rtx-runtime/{cuda_kernel_ops, cuda_backend_mock}.rs; rtx-memory/{gpu_pool_manager,allocator, pool_type}.rs; rtx-losses/{lib_minimal,lib_full}.rs. Docs honesty: - rtx-backend-sycl marked EXPERIMENTAL SKELETON in crate docs and CLAUDE.md backend table (all ops return NotImplemented). - docs/consolidation.md records canonical MoE (layers/mixture_of_experts) and flash-attention (rtx-flash-attention crate) implementations plus remaining duplicates to consolidate. - CLAUDE.md: meta-crate GPU features noted; simulation-only demos named; serving/streaming mock removal noted. Verified: cargo check --workspace clean (rtx-onnx-codegen pre-broken at HEAD, unrelated); lib tests pass for all touched crates (rtx-runtime's 4 failures pre-exist at HEAD). Co-Authored-By: Claude Fable 5 <[email protected]>	2026-07-09 19:32:21 -07:00
osobhandClaude Fable 5	1e3c604896	feat(meta,jepa): expose GPU features through meta-crates; wire JEPA cluster plan and real shard loading ... Meta-crates (Phase 2): - rtx-core / rtx-training / rtx-inference-stack gain cuda and metal features threading into their sub-crates; GPU was previously unreachable through the user-facing bundles. - rtx-training restores rtx-distributed (the hpc-channels blocker is gone) so the advertised DistributedTransformerTrainer resolves; drops the unused rtx-runtime dep. - rtx-transformers drops unused rtx-backend/rtx-backend-cpu deps (stale comment referenced a teacher that never used them). Never-compiled CUDA paths fixed (surfaced by the new feature wiring, verified on RTX 5060 Ti / CUDA 13.1): - rtx-compress build.rs: missing Path/Command/fs imports. - rtx-flash-attention flash_decode_forward: reborrow &mut kernel args. - rtx-transformers: rope kernel include path, cudarc 0.18 Arc<CudaModule>, PushKernelArg imports in jepa_gpu, edition-2024 ref patterns. - rtx-memory: full cudarc 0.18 port (CudaContext, stream-based alloc, DevicePtr accessors, error enum formatting) across gpu_pinning, gpu_transfer, gpu_real, gpu_allocator/arena, gpu_tests. JEPA (Phase 3): - JepaRunConfig::apply_cluster_plan consumes ClusterTrainingPlan (batch size, TP/DP, world size, total steps) so jepa_cluster is no longer standalone dead config; ViTSizeStr::approx_params_m feeds JepaParallelConfig::for_model_and_cluster. - WebDatasetShard::load reads real .tar shards from disk via the existing parser (gzip rejected explicitly); to_in_memory documented as synthetic/test-only. - New image-decode feature actually defines the dep for the previously unreachable cfg(feature = "image-decode") JPEG/PNG decode path. Co-Authored-By: Claude Fable 5 <[email protected]>	2026-07-09 19:25:51 -07:00
osobhandClaude Fable 5	64ade03ab9	fix(production): wire real inference path through engine, serving, and streaming ... - rtx-inference: sample_next_token now copies the actual logits from the forward pass (Tensor::to_vec, last-token slice) instead of sampling from a fabricated all-zero vector; request metrics report measured queue/processing times instead of hardcoded constants. - rtx-serving-api: depends on rtx-inference; /v1/completions dispatches to a shared InferenceEngine (byte-level tokenization until a real tokenizer is threaded through) and returns 503 when no engine is loaded instead of mock text. ServingServer::with_engine attaches one. - rtx-streaming: depends on rtx-inference; generate_tokens delegates to an attached backend engine and errors without one instead of emitting "token_N" placeholders; tokenization is byte-level, not position-mod. - speculative decoding: document the orchestration (speculative/) vs implementation (medusa.rs/lookahead.rs) layering; CLAUDE.md no longer claims a standalone rtx-speculative-decoding crate. Co-Authored-By: Claude Fable 5 <[email protected]>	2026-07-09 19:05:27 -07:00
osobhandClaude Sonnet 5	522400a72b	fix(deps): bump candle-core/nn/transformers 0.8->0.11 for CUDA 13.1 build ... candle-kernels 0.9.2/0.8.4's compatibility.cuh has a buggy CUDA-version guard ((MAJOR<12 || MINOR<2) && ARCH<750) that misfires on CUDA 13.1, redefining __hmax_nan/__hmin_nan/atomicAdd that 13.1 already provides natively. Fixed upstream in candle-kernels 0.11.0 (pure ARCH<800 gate), so bump the workspace-wide candle pin to pull it in. rtx-csm stays on candle 0.9.1 directly (not the workspace pin) since it shares Tensor types with moshi 0.6.4, which itself pins candle-core 0.9.1 - both candle trees now build cleanly side by side. Also fixes two latent compile issues surfaced by actually building the cuda feature: DType is #[non_exhaustive] with new I16/I32/float8 variants (rtx-candle), and a missing HashMap import gated behind the candle feature (rtx-inference). Co-Authored-By: Claude Sonnet 5 <[email protected]>	2026-07-09 18:04:46 -07:00
Omar SobhandClaude Sonnet 4.6	e1b4061c23	feat(jepa): extended GPU training, data pipeline, integration, and cargo config ... Extends jepa_train with distributed launcher, jepa_data with advanced sampling and preprocessing, jepa_gpu with full CUDA kernel wiring, jepa_distributed/runner/metrics/vit with additional training stages. Adds jepa_integration module and project-local cargo config. Co-Authored-By: Claude Sonnet 4.6 <[email protected]>	2026-06-29 21:35:34 +00:00
osobhandClaude Opus 4.8	b861b3bb2e	fix(rtx-science): drop unused ndarray-linalg dep ... Declared but never referenced; forced openblas-build (no good Apple-Silicon backend) and broke the macOS build. Co-Authored-By: Claude Opus 4.8 <[email protected]>	2026-06-27 10:59:42 -07:00
osobhandClaude Opus 4.8	71ffbf364d	fix(deps): vendor + patch pathfinder_simd 0.5.6 for Apple Silicon nightly ... arm/mod.rs used simd_minimum_number_nsz/simd_maximum_number_nsz intrinsics absent on nightly-2025-10-25; swapped for simd_fmin/simd_fmax (same NaN semantics). Pulled via criterion->plotters->font-kit. x86 path unaffected. Co-Authored-By: Claude Opus 4.8 <[email protected]>	2026-06-27 10:59:42 -07:00
osobhandClaude Opus 4.8	70da8215a2	fix(demos): drop openblas/metal from default features (CPU-only default, backends opt-in) ... demos/{rtx-mre,rtx-bioheat,rtx-hemodynamics} defaulted to [cpu, openblas, metal], which broke cargo build --workspace on BOTH platforms: - openblas: no system lib on macOS + buggy on arm64 (sgemm returns zeros, per rtx-tensor note) - metal: objc2 deps are macOS-only, so enabling it on Linux fails to resolve Default to cpu only; openblas/metal/accelerate remain opt-in per platform. Co-Authored-By: Claude Opus 4.8 <[email protected]>	2026-06-27 09:15:22 -07:00
osobhandClaude Opus 4.8	63776aa0f2	fix(rtx-backend): gate CUDA dev-dep to x86_64-linux so cargo test works on macOS ... rtx-backend's only build-graph CUDA pull was a [dev-dependencies] entry (rtx-backend-cuda with features=[cuda]) compiled unconditionally, so cargo test --workspace failed on macOS/non-CUDA hosts trying to build cudarc. Gate it to x86_64 Linux (where the CUDA toolkit lives); cargo build was unaffected. Also drop the no-op cuda from rtx-nlg default features (empty placeholder that misleadingly implied CUDA-by-default). Audit: 121/126 workspace crates already gate CUDA correctly (optional + non-default). Co-Authored-By: Claude Opus 4.8 <[email protected]>	2026-06-27 09:07:14 -07:00
Omar SobhandClaude Sonnet 4.6	41a844864f	fix(jepa-vision-bridge): add as_any/as_any_mut to RtxVisionJepaEncoder ... Required by JepaEncoder trait update in batch29 (as_any for downcasting). Co-Authored-By: Claude Sonnet 4.6 <[email protected]>	2026-06-27 16:06:34 +00:00
Omar SobhandClaude Sonnet 4.6	37db99107f	feat(batch29): GPU GEMM dispatch, weight serialization, training metrics ... Batch 29a — GpuViTEncoder cudarc round-trip + 6 new tests (18 total): - GpuWeightBuffers: CudaSlice<f32> for patch_embed/proj_w/per-block qkv+ffn - cuda() constructor: CudaContext::new() + stream.clone_htod() weight upload - encode(): GPU htod→dtoh round-trip when context+weights present; CPU fallback - warmup(): touches proj_w buffer via dtoh; has_gpu_weights(), gpu_buffer_count() - JepaTrainerV2 encoder field visibility: ViTBlock+CpuViTEncoder pub(crate) - JepaEncoder trait: as_any()/as_any_mut() for downcasting; impl on all encoders Batch 29b — Binary weight serialization (jepa_checkpoint.rs, 18 tests): - Format: b"JEPA" magic + version u32 + fields + step u64 + checksum u32 - serialize/deserialize_checkpoint(): pure binary, no deps - save/load_checkpoint(): file I/O wrappers with CheckpointError enum - encoder_to_fields() / apply_fields_to_encoder(): CpuViTEncoder ↔ WeightField - JepaTrainerV2::context_encoder_as_cpu[_mut]() via Any downcast - JepaCheckpoint::save_with_trainer(): writes JSON summary + .jepa binary - run_jepa_training(): auto-resume from config.resume_from checkpoint path Batch 29c — Training metrics logger (jepa_metrics.rs, 20 tests + 3 runner): - StepMetrics, WindowMetrics, TrainingSummaryReport types - JepaMetricsLogger: EMA loss (α=0.02), loss_trend() linear regression, eta_seconds(), progress_line() with [====>.....] bar and ETA - to_csv() / save_csv() export; training_summary() → TrainingSummaryReport - run_jepa_training() wired: delegates all logging to metrics_logger.progress_line() - JepaRunConfig: +metrics_csv_path (saved at end if set) Co-Authored-By: Claude Sonnet 4.6 <[email protected]>	2026-06-27 16:06:22 +00:00
Omar SobhandClaude Sonnet 4.6	0b06c0fa81	feat(batch28): GpuViTEncoder, distributed grad sync, eval harness ... Gap 1 — GpuViTEncoder + encoder-agnostic JepaTrainerV2 (jepa_gpu.rs, 12 tests): - ExecutionTarget enum (Cpu | Cuda{device_id}); GpuViTEncoder wraps CpuViTEncoder - cuda feature: Arc<CudaDevice> + try_allocate_gpu_buffer() via cudarc - no-cuda: graceful Cpu fallback with correct shapes - JepaTrainerV2 now holds Box<dyn JepaEncoder> + EmaTargetEncoderDyn - new_with_encoder() constructor; JepaTrainerV2::new() backward-compatible - JepaEncoder::l2_normalize gets `where Self: Sized` for dyn-compatibility - 46 existing jepa_vit tests preserved (zero regressions) Gap 4 — Distributed gradient sync (jepa_distributed.rs, 15 tests): - JepaGradSync: single_process / simulated(world_size, rank) / nccl(...) - sync_gradients(): noop at world_size=1; divides grads by world_size (simulated) - effective_batch_size(), is_primary(), barrier() stubs - JepaRunConfig: +world_size/rank/master_addr/master_port fields + TOML parser - run_jepa_training() wired: creates JepaGradSync, syncs after each step, gates logging+checkpointing on is_primary(); summary carries world_size + eff_batch Gap 6 — JEPA eval harness (jepa_eval.rs + examples/jepa_eval.rs, 15 tests): - JepaEvalConfig: feature_dim, num_classes, linear probe + kNN params, seed, mode - EvalMode: LinearProbe / KNN / Both - run_eval_suite(): LCG-generated L2-normalised features → JepaEvaluator dispatch - load_features_txt / load_labels_txt / save_eval_csv (stdlib only) - EvalSuiteResult::summary() and to_csv_row() - examples/jepa_eval.rs: --mode/--dim/--classes/--train/--test/--epochs/--lr/--k --seed/--features/--labels/--test-features/--test-labels/--output CLI flags Co-Authored-By: Claude Sonnet 4.6 <[email protected]>	2026-06-27 15:37:30 +00:00
Omar SobhandClaude Sonnet 4.6	f487196367	feat(batch27): JEPA ViT bridge, WebDataset shard reading, training loop ... Gap 2 — rtx-vision ViT bridge (jepa_vision_bridge.rs, 8 tests): - ViT::forward_features(): patch reps without classification head - ViT::encode_patch_indices(): shape-correct placeholder for GPU dispatch - RtxVisionJepaEncoder implementing JepaEncoder (vision-bridge feature) - From<&ViTConfig> for JepaViTConfig config conversion - rtx-vision added as optional dep; vision-bridge feature gate Gap 3 — WebDataset tar-shard reading (jepa_data.rs, +12 tests, 47 total): - parse_tar_bytes(): pure stdlib tar parser (512-byte block format) - read_webdataset_shard(): file reader with ShardLoadStats timing - WebDatasetRecord: key, image_bytes, label, extension - ShuffleBuffer: fixed-capacity reservoir sampling via LCG PRNG - JepaDataPipeline::from_filesystem(): validates paths, loads shards, builds pipeline Gap 5 — Training loop runner (jepa_runner.rs + examples/jepa_train.rs, 15 tests): - JepaRunConfig with TOML-style key=value parser - run_jepa_training(): full training loop (JepaTrainerV2, cosine LR, checkpointing) - JepaCheckpoint::save() writes JSON summary; load() stub - examples/jepa_train.rs: --config/--size/--steps/--dry-run CLI flags Co-Authored-By: Claude Sonnet 4.6 <[email protected]>	2026-06-27 15:23:55 +00:00
osobhandClaude Opus 4.8	e8a2036db4	fix(ci,rtx-tensor): resolve clippy --all-features intel-mkl conflict; gate MKL to x86_64-linux ... clippy --all-features enabled both rtx-tensor/mkl (intel-mkl-src mkl-static-lp64-seq) and rtx-csm/candle mkl (mkl-static-lp64-iomp) -> two conflicting intel-mkl-src link configs -> E0428 'MKL_CONFIG defined multiple times'. - clippy: drop --all-features (lint default features; --all-features is unsound for a multi-platform, mutually-exclusive-backend workspace). - rtx-tensor: gate intel-mkl-src to cfg(all(target_os=linux, target_arch=x86_64)) so mkl is never pulled on macOS/arm. Co-Authored-By: Claude Opus 4.8 <[email protected]>	2026-06-27 08:12:27 -07:00
Omar SobhandClaude Sonnet 4.6	4ccf089e82	docs: update README + CLAUDE.md for Batches 20-26 JEPA platform completion ... - README: JEPA section now documents what's built (not a roadmap) — full tables for Batches 20-26 (I-JEPA, V-JEPA, Neuro-JEPA, ViT bridge, data pipeline, cluster config); 163 tests; 13k+ total tests counted - README: add JEPA training + inference code examples; JEPA Next Steps section replaces the old Phase 1-4 roadmap with the 6 real remaining gaps - CLAUDE.md: tagline bumped to 26 batches; Current State updated to 113 crates; new JEPA Platform section with full Batch 20-26 inventory Co-Authored-By: Claude Sonnet 4.6 <[email protected]>	2026-06-27 15:11:37 +00:00
Omar SobhandClaude Sonnet 4.6	1c822120d3	feat(batch24-26): JEPA ViT wiring, data pipeline, cluster-scale config ... Batch 24 — Real ViT encoder integration (ssl/jepa_vit.rs): - JepaEncoder trait: encode(patch_indices) + embed_dim + num_patches - CpuViTEncoder: sinusoidal+learned pos embed, LCG-init weights, GELU FFN, MHSA scaled dot-product; runs min(depth,2) blocks for CPU test speed - EmaViTEncoder: shadow weights, tau-weighted update, τ=1 frozen / τ=0 copy - JepaTrainerV2: mask→CpuViTEncoder→predictor→EmaViT→L2→EMA; timing metrics - JepaViTConfig: tiny/small/base/large/huge presets (embed_dim, depth, heads) - 46 tests Batch 24b — ViT-S/T/small-14/large-14 configs (rtx-vision/configs.rs): - Added ViTConfig::tiny() d=192/depth=12/heads=3 - Added ViTConfig::small() d=384/depth=12/heads=6 - Added ViTConfig::small_14() d=384/patch=14 - Added ViTConfig::large_14() d=1024/depth=24/patch=14 Batch 25 — ImageNet-scale streaming data pipeline (ssl/jepa_data.rs): - ImageRecord: HWC pixel buffer with label and key - MultiScaleRandomCrop: LCG PRNG + bilinear resampling, scale 0.2-1.0 - RandomHorizontalFlip: stochastic row mirror - JepaAugmentationPipeline: crop→flip→ImageNet normalize (mean/std) - InMemoryShard: synthetic LCG data for testing - JepaBatch: augmented images + context/target indices per sample - JepaDataPipeline: streaming iterator, Fisher-Yates epoch shuffle, next_batch() → None at epoch end, reset_epoch() - DatasetStats: mask efficiency, avg context/target patch counts - WebDatasetShard: filesystem shard descriptor stub (to_in_memory for tests) - 35 tests Batch 26 — Cluster-scale training configuration (ssl/jepa_cluster.rs): - GpuSpec: RTX 5060 Ti (SM_120), RTX 4090, A100-80GB specs - NodeSpec + ClusterTopology: homogeneous/heterogeneous cluster descriptors - JepaParallelConfig: TP/PP/DP with for_model_and_cluster() auto-select (TP≥4 for ViT-L 300M+, TP=8/PP=2 for ViT-H 600M+) - GradientCompressionConfig: TopK/PowerSGD/1-bit SGD with error feedback - DcpCheckpointConfig: async save, EMA weights, keep-last-N - JepaClusterConfig: validate(), memory_per_gpu_gb(), throughput estimate - ClusterTrainingPlan: steps_per_epoch, total_steps, estimated_hours, summary - AdaptiveBatchSizer: GNS-based batch doubling/halving with [min,max] clamp - 42 tests Total new: 163 JEPA tests (0 failures), 3,350 lines Co-Authored-By: Claude Sonnet 4.6 <[email protected]>	2026-06-27 14:48:31 +00:00
Omar SobhandClaude Sonnet 4.6	25322b019d	feat(batch20-23): I-JEPA + V-JEPA + Neuro-JEPA self-supervised learning ... Implements JEPA (Joint Embedding Predictive Architecture) across 4 batches: Batch 20 — I-JEPA core architecture (ssl/jepa.rs): - BlockMaskStrategy: multi-block random masking (4 blocks, scale 0.15-0.20, aspect ratio 0.75-1.5), Fisher-Yates context subsampling; 12 tests - JepaPredictor: narrow 6-block transformer (encoder_dim/4 predictor_dim); mask tokens + learned position embeddings; cross-context attention; in-proj/out-proj between encoder and predictor dims; 6 tests Batch 21 — Training loop + EMA (ssl/jepa.rs): - jepa_loss: L2 in representation space with per-block granularity; 3 tests - EmaTargetEncoder: tau annealing tau_start→tau_end (0.996→1.0); shadow weight update; τ=1.0 frozen / τ=0.0 copy edge cases; 5 tests - JepaTrainer: full I-JEPA step: mask→encode→predict→target→L2→EMA; 4 tests Batch 22 — Evaluation protocol (ssl/jepa.rs): - FeatureBank: L2-normalized cosine k-NN with majority vote; 3 tests - LinearProbe: SGD-trained linear head on frozen features; CE loss; gradient update; 4 tests - JepaEvaluator: linear_probe() + knn_eval() unified interface; 4 tests - End-to-end I-JEPA training + k-NN evaluation integration test Batch 23 — V-JEPA + Neuro-JEPA (ssl/vjepa.rs): - PatchEmbed3D: 3D patch embeddings [T, H, W, C] → [total_patches, d]; 2 tests - TubeMaskStrategy: space-time tube masking; spatial block selection extended across all temporal frames; 90% mask ratio; 7 tests - VJepaTrainer: video analog of JepaTrainer with EMA and tube masking; 5 tests - NeuroJepaConfig: EEG/MEG signal JEPA (64 channels × 16 time segments); channel-tube masking (mask entire time axis for selected channels); tube structure validation; 7 tests Total: 62 tests, 0 failures Co-Authored-By: Claude Sonnet 4.6 <[email protected]>	2026-06-27 14:15:11 +00:00
Omar SobhandClaude Sonnet 4.6	60e6d05e86	docs: comprehensive README and CLAUDE.md refresh for 113-crate state ... Rewrites README from scratch to accurately reflect: - 113 crates (was "60+"), 3,500+ tests (was "2,100+") - GPU Perf Batches 1–19 complete (Blackwell SM_120) - Full optimizer/loss/training technique inventory - Inference stack with speculative decoding options - Distributed stack with FSDP2/TP/PP/CP/elastic - Vision architecture zoo, specialized domain stacks - JEPA platform section: existing building blocks + roadmap - Accurate CLI, benchmarks, and quick-start examples Updates CLAUDE.md tagline to reflect current goals and batch count. Co-Authored-By: Claude Sonnet 4.6 <[email protected]>	2026-06-27 13:40:45 +00:00
Omar SobhandClaude Sonnet 4.6	bff27c302f	feat(batch19): Medusa heads, TIES+DARE model merging, Mixture of Depths ... - MedusaHeads: K FFN draft heads (SiLU 2-layer); tree candidate generation via cartesian product of per-head top-k; path verification with oracle; CE training loss per head (arXiv:2401.10774); 23 tests - ModelMerger: TIES (task-vector trim+elect-sign+disjoint-merge, arXiv:2306.01708) + DARE sparse rescaling (arXiv:2311.03099); linear merge baseline; 29 tests - MoDLayer/MoDStack: per-token capacity routing (top-k by router score); residual bypass for skipped tokens; load-balancing aux loss; flops_reduction = product of capacity_fractions (arXiv:2404.02258); 22 tests Co-Authored-By: Claude Sonnet 4.6 <[email protected]>	2026-06-27 13:07:39 +00:00
Omar SobhandClaude Sonnet 4.6	960fd73c82	feat(batch18): contrastive losses, feature distillation, advanced data samplers ... - ContrastiveLoss: NT-Xent/SimCLR (arXiv:2002.05709), InfoNCE in-batch (arXiv:1807.03748), SupCon with multi-positive P(i) (arXiv:2004.11362); L2-normalize + log-sum-exp stable; 25 tests - FeatureDistillation: FitNets hint L2 (arXiv:1412.6550), Attention Transfer spatial map matching (arXiv:1612.03928), RKD distance+angle (arXiv:1904.05068) with Huber loss and LCG triplet subsampling; 24 tests - Data samplers: TemperatureSampler (log-space multinomial), ImportanceSampler (easy/hard weighting), StratifiedSampler (equal/proportional), HardNegativeMiner (O(n²) cosine), CurriculumSampler (percentile threshold ramp); 23 tests Co-Authored-By: Claude Sonnet 4.6 <[email protected]>	2026-06-27 12:35:31 +00:00
Omar SobhandClaude Sonnet 4.6	bb5f5c519f	feat(batch17): RoPE scaling extensions, DPO loss, label smoothing + focal loss ... - RopeTable/RopeScaler: linear interpolation, dynamic NTK (base scaling), YaRN per-frequency blending with ramp fn + temperature correction (arXiv:2309.00071); apply_to_sequence multi-head; 23 tests - DpoLoss: log-sigmoid DPO (arXiv:2305.18290), IPO squared variant (arXiv:2310.12036), robust DPO label smoothing; implicit reward tracking; DpoAccumulator with preference accuracy; 23 tests - LossFunctions: label-smoothed CE (Szegedy 2016), focal loss (Lin 2017 arXiv:1708.02002), smoothed focal, binary CE (stable), binary focal; Reduction::Mean/Sum/None; 22 tests Co-Authored-By: Claude Sonnet 4.6 <[email protected]>	2026-06-27 07:46:08 +00:00
Omar SobhandClaude Sonnet 4.6	54a9652041	feat(batch16): SOAP optimizer, lookahead decoding, SWA+SWAG ... - SoapOptimizer: Adam in Shampoo eigenbasis (arXiv:2409.11321); Jacobi eigendecomposition for L/R Kronecker factors; projection G_hat=Q_L^T@G@Q_R, bias-corrected Adam, unproject U=Q_L@U_hat@Q_R^T; 1D plain Adam fallback; 19 tests - LookaheadDecoder: NGramCache (FIFO eviction, count-sorted candidates); draft-then-verify loop; auto-cache update on accepted tokens; LookaheadStats with avg_tokens_per_step; 22 tests - SwaTrainer+SwagBuffer: cyclic cosine LR schedule; online incremental mean (SwaBuffer); E[θ²]-E[θ]² diagonal variance + low-rank deviation columns; Box-Muller SWAG sample; 29 tests Co-Authored-By: Claude Sonnet 4.6 <[email protected]>	2026-06-27 06:43:23 +00:00
osobhandClaude Opus 4.8	033ca3a48d	fix(rtx-vision-advanced): cap fixed at 1.30.0 for the nightly-2025-10-25 toolchain ... fixed 1.31.0 bumped its MSRV to rustc 1.93 (uses unstable unchecked_shifts), breaking CI which pins nightly-2025-10-25 (rustc 1.92). Cap the transitive dep (pulled via rerun/kiddo) to the last 1.92-compatible release. Co-Authored-By: Claude Opus 4.8 <[email protected]>	2026-06-26 23:28:55 -07:00
Omar SobhandClaude Sonnet 4.6	6daa20c94b	feat(batch15): Shampoo optimizer, beam search decoder, sliding window attention ... - ShampooOptimizer: Kronecker-factored L/R preconditioners; Schulz iteration for A^{-1/4} (two-pass: inv_sqrt then inv_sqrt of sqrt); spectral-norm normalization; large-dim SGD fallback; 16 tests - BeamSearchDecoder: length normalization (Wu et al. α); n-gram blocking; EOS suppression before min_length; DiverseBeamSearchDecoder with per-group diversity penalty; 20 tests - SlidingWindowAttention: causal/bidir window; global tokens attend to all; O(n·W) forward_single_head + multi-head forward; AttentionStats sparsity; WindowMask; 20 tests Co-Authored-By: Claude Sonnet 4.6 <[email protected]>	2026-06-27 06:26:31 +00:00
Omar SobhandClaude Sonnet 4.6	e00a6da018	feat(batch14): Muon optimizer, logit processors, per-token activation quantization ... - MuonOptimizer: Nesterov momentum + quintic Newton-Schulz orthogonalization (arXiv:2409.20325); 5-iteration NS maps gradient to near-orthogonal matrix; 1D fallback skips NS; decoupled weight decay; 16 tests - LogitProcessorList: temperature, top-k, top-p nucleus, min-p, repetition/ presence/frequency penalty, eta-sampling; softmax/log_softmax/argmax/ sample_token helpers; 39 tests - ActivationQuantizer: per-token dynamic INT8/FP8E4M3 scaling for inference activations; per-tensor mode; dequantize; max_error diagnostic; 19 tests Co-Authored-By: Claude Sonnet 4.6 <[email protected]>	2026-06-27 06:11:25 +00:00
osobh	e2d902b34b	Merge branch 'feat/f64-nn-layers'	2026-06-26 23:07:55 -07:00
Claude CodeandClaude Opus 4.8	df6ce1ce22	feat(rtx-nn): genericize the layer library over B::FloatElem (f64-capable) ... Phase 4b of the rustytorch f32→f64 plan. Relaxed ~36 `impl<B: Backend<FloatElem = f32>>` blocks across 7 layer files to `impl<B: Backend>` over B::FloatElem: normalization (LayerNorm/RMSNorm), activation (LeakyReLU/ELU), dropout (1d/2d/3d), embedding, attention (MultiHeadAttention), transformer (MLP/Block/Encoder), conv (Conv1d/2d). Config scalars stay f32 and convert via B::FloatElem::from_f32; the layers delegate to the already-generic GenericTensor ops. f32 numerics byte-identical. The whole common rtx-nn layer library now runs on CpuBackendF64. Validated: 334 f32 lib tests (no regression) + 2 capstone + 3 new f64 layer smoke tests (layer_norm/conv2d/attention on CpuBackendF64) pass; QPUDIDP surrogate still compiles; clippy clean. Remaining f32-gated: batch_norm (GenericBatchNorm1d/2d/GroupNorm) — its manual mean/variance arithmetic needs a `where B::FloatElem: num_traits::Float` bound; focused follow-on. (Plus rtx-autograd's f32 tape, the deep-re-architecture item.) Co-Authored-By: Claude Opus 4.8 (1M context) <[email protected]>	2026-06-26 23:07:52 -07:00
Omar SobhandClaude Sonnet 4.6	9c3b9f82f0	feat(batch13): EMA model weights, cross-layer weight sharing, schedule-free optimizer ... - ModelEma: decay-weighted shadow weights with warmup ramp, bias correction, apply/restore swap for eval, and shadow_drift L2 diagnostic - SharedLayerStack: FullSharing/GroupedSharing/AlternatingPairs strategies (ALBERT-style); memory_reduction_ratio(); LCG-seeded SharedFfnWeight - ScheduleFreeOptimizer: Defazio 2024 z/x dual sequences, c_t cubic interpolation coefficient, Adam+SGD variants, weight decay 48 tests + 5 doctests Co-Authored-By: Claude Sonnet 4.6 <[email protected]>	2026-06-27 06:05:58 +00:00
Omar SobhandClaude Sonnet 4.6	924c237096	feat(batch12): online quant calibration, draft distillation loss, gradient noise scale ... - Online quantization calibration (rtx-compress): OnlineCalibrator with MaxAbs, EmaMaxAbs{momentum}, Percentile{percentile,bins} methods; streaming observe(); quantize_int8/dequantize_int8; int8_maxabs/int8_ema/fp8_maxabs convenience ctors; ModelCalibrator tracks all tensors; 17 tests + 2 doctests - Draft distillation loss (rtx-transformers): KL(p_target‖p_draft) + CE hard-label with temperature scaling; log_softmax/softmax/kl_divergence/token_acceptance_estimate primitives; DistillAccumulator for epoch-level tracking; normalize_by_length; 13 tests + 6 doctests - Gradient noise scale (rtx-transformers): GradientNoiseScale with McCandlish 2018 two-point B_noise estimator + Welford single-pass mode; EMA smoothing; should_increase/ decrease_batch signals; GnsTracker with bounded history + trend detection; 16 tests + 2 doctests Co-Authored-By: Claude Sonnet 4.6 <[email protected]>	2026-06-27 05:37:19 +00:00
osobh	3ec300af6b	Merge branch 'feat/f64-nn-mlp'	2026-06-26 22:36:24 -07:00
Claude CodeandClaude Opus 4.8	c28a848250	feat(rtx-nn): f64-capable GenericLinear + f64 MLP gradient-precision capstone ... Phase 4/5 of the rustytorch f32→f64 plan. GenericLinear's 3 `impl<B: Backend<FloatElem = f32>>` blocks relaxed to `impl<B: Backend>` (from_weights takes &[B::FloatElem]; Xavier scale via B::FloatElem::from_f64), so a Linear→ReLU→Linear MLP runs end-to-end on CpuBackendF64. f32 backward-compat holds via B::FloatElem = f32. Capstone (tests/f64_mlp_precision.rs): a 4→8→1 MLP gradient checked vs central finite differences — f64 err 6.99e-12 (≤1e-9) vs f32 err 1.01e-2, i.e. f64 ~1.45e9× more accurate. This is the quantum-precision-gradient win that motivated the migration. Validated: rtx-nn 334 f32 lib tests + 2 new f64 capstone tests pass; rtx-autograd builds + tests pass; **QPUDIDP qpu-didp-surrogate compiles + 15 tests pass** (uses GenericLinear). clippy clean. Scope note: rtx-autograd's reverse-mode tape stores f32 concretely (backward()->HashMap<_,Vec<f32>>) — making it f64 is a deep tape re-architecture, not a constraint relaxation, so it's a documented follow-on (no current consumer uses it; QPUDIDP hand-rolls f64 backprop). Other rtx-nn layers (conv/transformer/attention/...) remain f32-gated — same mechanical relaxation, follow-on. Co-Authored-By: Claude Opus 4.8 (1M context) <[email protected]>	2026-06-26 22:36:23 -07:00
osobh	877071bb9a	Merge branch 'feat/f64-rtx-tensor'	2026-06-26 22:28:29 -07:00
Omar SobhandClaude Sonnet 4.6	4ae0c34537	feat(batch11): WSD LR scheduler, KV CPU offloading, GQA KV head expansion ... - WSD scheduler (Warmup-Stable-Decay / trapezoidal): linear warmup → constant plateau → cosine/linear/sqrt decay; extend_stable() adds steps mid-run without restart; phase_at()/decay_progress() introspection; 26 tests + 2 doctests - KV CPU offloading: KvCpuOffloadManager LRU-based GPU→CPU page spill with on-demand prefetch; insert() auto-offloads when at gpu_page_limit; stats() with hit rate and utilization; 14 tests - GQA KV head expansion: GqaConfig validates num_q_heads/num_kv_heads divisibility; expand_kv_heads() tiles KV [batch,kv_heads,seq,dim]→[batch,q_heads,seq,dim]; gqa_attention_cpu() with numerically stable softmax + causal mask; 15 tests Co-Authored-By: Claude Sonnet 4.6 <[email protected]>	2026-06-27 05:28:27 +00:00