[NVIDIA] Fix DeepSeek V4 mHC warmup coverage #1

2026-07-30T09:39:41Z

karylab_agents commented

2026-07-30 09:39:41 +00:00

Summary

expose the NVIDIA DeepSeek V4 decoder hc_pre and hc_post entry points to the warmup detector
warm the first-layer broadcast mHC path before serving
compile the reachable n_splits token-size keys instead of allowing TileLang JIT compilation during inference
add focused unit coverage for detector guards, split-key selection, and broadcast/non-broadcast warmup paths

Why

On the dual-GB10 TP=2 deployment, mHC TileLang kernels were still compiling during live requests. Repeated mhc_pre_*, mhc_post_tilelang, and hc_head_fuse_tilelang compilation could exhaust shared-memory broadcast resources and wedge workers until the RPC timeout. The existing warmup detector silently skipped NVIDIA decoder layers because those layers did not expose hc_pre / hc_post, and the broadcast path was not covered.

Provenance

This is a clean backport of the still-open upstream vLLM PR:

https://github.com/vllm-project/vllm/pull/49707
source commits 32389a6b and 21ef130e

The staged patch ID was verified to be identical to the upstream two-commit patch. It is intentionally separate from the larger SM12x/GB10 preview branch so this root-cause fix can be reviewed independently.

Validation

git diff --check origin/main...HEAD: passed
patch-ID comparison against upstream PR #49707: identical
branch is one commit ahead of the current origin/main with no divergence
local CUDA unit tests were not run: this workstation has no Linux/CUDA vLLM test environment or uv
target validation still required: rebuild the GB10 image, verify startup warmup, and run a long multi-turn TP=2 workload while confirming no mHC TileLang compilation appears after the API becomes ready

AI assistance

Codex assisted with log diagnosis, upstream duplicate-work search, backport preparation, and static verification. A human maintainer should review and own the final merge and GB10 hardware validation.

## Summary - expose the NVIDIA DeepSeek V4 decoder `hc_pre` and `hc_post` entry points to the warmup detector - warm the first-layer broadcast mHC path before serving - compile the reachable `n_splits` token-size keys instead of allowing TileLang JIT compilation during inference - add focused unit coverage for detector guards, split-key selection, and broadcast/non-broadcast warmup paths ## Why On the dual-GB10 TP=2 deployment, mHC TileLang kernels were still compiling during live requests. Repeated `mhc_pre_*`, `mhc_post_tilelang`, and `hc_head_fuse_tilelang` compilation could exhaust shared-memory broadcast resources and wedge workers until the RPC timeout. The existing warmup detector silently skipped NVIDIA decoder layers because those layers did not expose `hc_pre` / `hc_post`, and the broadcast path was not covered. ## Provenance This is a clean backport of the still-open upstream vLLM PR: - https://github.com/vllm-project/vllm/pull/49707 - source commits `32389a6b` and `21ef130e` The staged patch ID was verified to be identical to the upstream two-commit patch. It is intentionally separate from the larger SM12x/GB10 preview branch so this root-cause fix can be reviewed independently. ## Validation - `git diff --check origin/main...HEAD`: passed - patch-ID comparison against upstream PR #49707: identical - branch is one commit ahead of the current `origin/main` with no divergence - local CUDA unit tests were not run: this workstation has no Linux/CUDA vLLM test environment or `uv` - target validation still required: rebuild the GB10 image, verify startup warmup, and run a long multi-turn TP=2 workload while confirming no mHC TileLang compilation appears after the API becomes ready ## AI assistance Codex assisted with log diagnosis, upstream duplicate-work search, backport preparation, and static verification. A human maintainer should review and own the final merge and GB10 hardware validation.

karylab commented

2026-07-30 09:40:57 +00:00

發到karylab/gb10分支

karylab_agents changed target branch from main to karylab/gb10

2026-07-30 09:44:56 +00:00

karylab_agents added 1 commit 2026-07-30 09:44:56 +00:00

Fix DeepSeek V4 NVIDIA mHC warmup coverage

New PR Bot / update-description (pull_request_target) Successful in 3s

Details

New PR Bot / reminder-comment (pull_request_target) Failing after 2s

Details

pre-commit / pre-run-check (pull_request) Canceled after 0s

Details

pre-commit / pre-commit (pull_request) Canceled after 0s

Details

d789590bb6

Expose the NVIDIA decoder mHC entry points to the warmup detector, warm the broadcast path, and compile the reachable split-key token sizes before serving.

Backport of the fix from vllm-project/vllm#49707 (commits 32389a6b and 21ef130e).

karylab merged commit 13188f3315 into karylab/gb10

2026-07-30 09:45:52 +00:00

Sign in to join this conversation.

2 Participants

Notifications

Due Date

No due date set.

Dependencies

No dependencies set.

Reference: Karylab-cklius/vllm#1

Allow edits from maintainers