Skip to content

feat: DeepSeek-V4-Flash integration + Blackwell (sm120) bring-up #153

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Draft
drunkcoding wants to merge 96 commits into
base: main
Choose a base branch
from
Draft

feat: DeepSeek-V4-Flash integration + Blackwell (sm120) bring-up #153

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
96 commits
Select commit Hold shift + click to select a range
02bba41
add gitignore
May 12, 2026
dc46551
test: add shared kernel test infrastructure (conftest + package init)
May 21, 2026
a395435
feat: add per-head fused QK RMSNorm Triton kernel for V4 MLA
May 21, 2026
7033852
feat: add native CUDA fused QK RMSNorm for V4 MLA (Path A)
May 21, 2026
3cf542a
feat(kernels): consolidate v4 kernels into batchgen_kernels
May 24, 2026
e0bb7dc
feat(kernels): v4 DSA attention kernel improvements
May 24, 2026
5d9e63e
test: add v4 kernel test suite
May 24, 2026
049b0a3
feat(model): V4 Flash model scaffolding + HC kernel wiring
May 24, 2026
b31d003
chore: docker base image update + vllm entry points doc
May 24, 2026
6994ad6
fix(kv-cache): correct V4 KV profile to 584 bytes/token
May 24, 2026
25998e2
feat(model): wire V4 gate routing, MoE activation, and attention decode
May 24, 2026
51c3127
test: add V4 wiring correctness tests and fix HC references
May 24, 2026
3afd3ae
fix(v4flash): make DP-decode MoE collective symmetric and correct
Jun 1, 2026
9628082
fix(v4flash): wire c4 indexer DP-gather + functional compressor for d…
Jun 2, 2026
d3d129a
fix(v4flash): pass compressed seqlen to c4 sparse indexer (fix decode…
Jun 2, 2026
cdc5c1a
fix(v4flash): gather decode results from global_batch, not local maps
Jun 2, 2026
92b4fb5
feat(v4flash): add sm120 Triton kernels for grouped MoE and sparse ML…
Jun 4, 2026
13574ff
feat(v4flash): add DeepSeek-V4 single KV pool and decode KV coordinator
Jun 4, 2026
5bd3250
feat(v4flash): add sparse MLA decode adapter, torch reference, and pr…
Jun 4, 2026
e86e584
feat(v4flash): wire sm120 decode path with grouped MoE, persistent ex…
Jun 4, 2026
f462863
feat(v4flash): add checkpoint metadata loader and assets packaging
Jun 4, 2026
7945f3c
fix(v4flash): register tokenizer discovery for V4-Flash
Jun 4, 2026
1a62675
chore(models): align v3/glm5/minimax/kimi config with shared decode c…
Jun 4, 2026
6b0e660
chore(kv-cache): update host KV manager config
Jun 4, 2026
bea1c26
feat(server): wire decode-timing step_done and worker/process management
Jun 4, 2026
4be62c2
feat(core): expose get_tensor and weight-buffer support for persisten…
Jun 4, 2026
9bbb050
build(kernels): update sm120 kernel build, op builder, and docker image
Jun 4, 2026
d7a1874
test(v4flash): add V4 decode integration, KV coordinator, and compres…
Jun 4, 2026
4d765cf
test(e2e): add serving benchmark and cross-framework MMLU-Pro harness
Jun 4, 2026
6628d51
chore(tools): add V4 repro, divtrace, and MoE analysis scripts
Jun 4, 2026
9ce9af0
chore: ignore generated batch storage artifacts
Jun 6, 2026
41ca4bf
feat(v4flash): support prepacked multi-seq prefill slicing
Jun 6, 2026
de030c4
style(server): normalize worker main loop indentation
Jun 6, 2026
a081227
feat(kernels): retarget sm120 JIT extensions for Blackwell
Jun 6, 2026
0d6da90
feat(kv-cache): isolate host KV shared memory names
Jun 6, 2026
faa39e2
feat(server): expose prepack and memory tuning flags
Jun 6, 2026
062b947
fix(parameter-server): support V4 remote weight serving
Jun 6, 2026
a67fd17
feat(v4flash): add grouped MXFP4 MoE decode path
Jun 6, 2026
2ccdfda
perf(v4flash-attn): vectorize sparse index physicalization
Jun 6, 2026
3b4e323
test(e2e): add V4 MMLU-Pro batch harness
Jun 6, 2026
acb143c
perf(v4flash-attn): batch C128 decode metadata updates
Jun 6, 2026
39e96a8
perf(v4flash): consolidate to 3D grouped MXFP4 MoE; sm120 CUDA 12.9 b…
Jun 9, 2026
642f698
Merge pull request #251 from batchgen-project/fix/v4flash-tokenizer-a…
drunkcoding Jun 9, 2026
0f6c393
Merge pull request #253 from batchgen-project/perf/v4flash-sm120-decode
drunkcoding Jun 9, 2026
7f69ce1
feat(ckpt-converter): add load_rank_shard_tensors for selective rank-…
Jun 15, 2026
3aef181
fix(v4flash): correct fp8 cast in act_quant and add shared-expert swi…
Jun 15, 2026
4ae7285
feat(v4flash): add sparse prefill attention ported from official refe…
Jun 15, 2026
2fd7f3d
feat(v4flash): wire sparse prefill routing, opt-in QAT linear, and fi…
Jun 15, 2026
7405257
fix(v4flash): use compression-aware rope cache for sparse and compres…
Jun 15, 2026
a987091
feat(v4flash-attn): add tensor-dump diagnostics and guard page-table …
Jun 15, 2026
fcd289a
fix(server): make NCCL timeout configurable for long-sequence offload…
Jun 15, 2026
722898d
fix(v4flash): keep prefill experts streamed and load rank-local vocab…
Jun 15, 2026
9d46f71
fix(kv-cache): charge V4 KV pools in compressed token space to preven…
Jun 15, 2026
3d408fa
fix(v4flash): join vocab-parallel collectives on 0-seq ranks and use …
Jun 15, 2026
944a57e
chore(tools): add V4 MMLU eval, divtrace, and sanity scripts
Jun 15, 2026
2ab2c53
docs: add DeepSeek-V4-Flash Blackwell bring-up handoff notes
Jun 15, 2026
12e2e2a
fix(v4flash): make ON_HOLD eviction V4-coordinator-aware to avoid crash
Jun 15, 2026
3db3385
fix(v4flash): collectively re-init GPU KV coordinator on watermark re…
Jun 15, 2026
8845d37
docs: record watermark on-hold/re-prefill crash fixes and QAT char-ex…
Jun 15, 2026
bce0b40
docs: localize residual decode drift to non-QAT grouped MoE kernel
Jun 15, 2026
986331e
test(v4flash): add grouped MoE kernel vs per-expert parity test (xfail)
Jun 15, 2026
fdca802
feat(v4flash): add QAT-faithful grouped MoE decode kernel (BATCHGEN_V…
Jun 15, 2026
22c8686
docs: record QAT MoE kernel result and residual #2 (decode attention/…
Jun 15, 2026
e72f1d5
docs: narrow residual #2 (rule out SM120 kernel + rope mismatch)
Jun 15, 2026
ee077d2
docs: refute fp8 KV, localize residual #2 to sparse indexer scoring
Jun 15, 2026
5d61f23
docs: validate indexer missing-relu changes ~30% of topk KV selection
Jun 15, 2026
42fd462
fix(v4flash): apply per-head relu in sparse indexer score to match of…
Jun 15, 2026
cba7f54
fix(v4flash): fp4-fake-quant indexer q and k to match official QAT
Jun 15, 2026
388da19
docs: record indexer relu (major) + fp4 (marginal) char-exact results
Jun 15, 2026
30ffb98
docs: root-cause residual to inherent fp rounding (router + sink refu…
Jun 15, 2026
bacba50
build(docker): add GPU_ARCH=hopper support to batchgen image
Jun 16, 2026
0143b23
chore(v4flash): drop stale handoff/diagnostic docs and tooling
Jun 23, 2026
a579e7f
perf(v4flash): vectorize sm120 decode index builders
Jun 23, 2026
813c791
perf(v4flash): warmup, pin autotune, and guard cache geometry for sm1…
Jun 23, 2026
c911974
perf(v4flash): vectorize fp8 KV-cache pack (7-tile loop -> batched)
Jun 23, 2026
7f10e5e
test(v4flash): fix stale wkv refs and add eager<->kernel compressor p…
Jun 23, 2026
4a594b3
feat(v4flash): sm-aware indexer quant (MXFP4 on sm120, FP8 on sm90)
Jun 23, 2026
9c558bf
feat(v4flash): gate grouped MXFP4 MoE off below sm120
Jun 23, 2026
cdfc8a8
test(v4flash): skip-guard MXFP4-only tests below sm120 + add fp8/disp…
Jun 23, 2026
95a0c3e
test(v4flash): fix stale wkv/world_size refs + tilelang guards in dec…
Jun 23, 2026
e19e464
docs(docker): add H20 rebuild+launch runbook for full V4-Flash run
Jun 23, 2026
ba22d80
build(docker): add CN-mirror build args + harden GitHub clones
Jun 24, 2026
cc90d02
build(docker): wire CN-mirror build args into H20 runbook script
Jun 24, 2026
e4d84dc
fix(v4flash): make DP-decode entry collectives rank-safe
Jun 25, 2026
8dd373a
feat(v4flash): add BATCHGEN_DECODE_DEADLOCK_TRACE rank markers
Jun 25, 2026
e6a4988
perf(v4flash): batch per-expert host sync in _run_owned_experts
Jun 25, 2026
5000c7e
build(docker): persist torch JIT cache + fix memlock for V4 H20 run
Jun 25, 2026
ea79238
build(deps): bump 7 deps to patch Dependabot vulnerabilities
Jun 25, 2026
d02be92
feat(v4flash): add sm120 grouped-MoE kernels (mega3/mega/graph/ragged…
Jul 4, 2026
9bf47d8
feat(v4flash): wire mega3 MoE + resident-experts/EP-offload serving path
Jul 4, 2026
7c25580
fix(v4flash): size RoPE caches to original_seq_len, grow on demand
Jul 4, 2026
c8cce99
fix(v4flash): GPU-KV admission backpressure for prefill and decode
Jul 4, 2026
14d4593
test(v4flash): numerics parity, arch gate, and MoE wiring parity tests
Jul 4, 2026
ab125aa
build(docker): update H20 runbook for sm120 kernel integration
Jul 4, 2026
f824046
docs(v4flash): 4xRTX6000-Pro setup guide, kernel/EP notes, autoresear…
Jul 4, 2026
acf8469
feat(planner): env overrides for phase/module micro-batching knobs
Jul 4, 2026
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
17 changes: 17 additions & 0 deletions .gitignore
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -242,6 +242,7 @@ cython_debug/
#.idea/

# Server runtime storage (batch metadata, uploaded files, outputs)
batchgen/storage/
batchgen/storage/batches/
batchgen/storage/files/
batchgen/storage/files_meta/
Expand All @@ -264,6 +265,22 @@ memory/
session-state/
*.session.jsonl
task_queue.jsonl
.sisyphus
CLAUDE.md

# Sisyphus / OhMyOpenCode (omo) agent docs — keep local, never commit
AGENTS.md
GEMINI.md
.opencode/
.omo/
omo-*.md
*.omo.md
opencode.json

# Agent handoff / session docs — keep local
HANDOFF.md
HANDOFF-*.md

# Kernel-development tree (lives in Andrewxu313/batchgen_kernel_dev, not here)
batchgen_kernel_dev/
benchmarks/
Loading