From ddbbfd29e52d602326c258dc68f785e855d2324c Mon Sep 17 00:00:00 2001 From: Hongsheng Liu Date: Thu, 6 Aug 2026 10:54:00 +0800 Subject: [PATCH] docs: reorganize design navigation Signed-off-by: Hongsheng Liu --- docs/.nav.yml | 68 ++- docs/design/blackwell_production_support.md | 618 ++++++++++++++++++++ docs/design/diffusion_ops.md | 231 ++++++++ docs/design/index.md | 59 +- docs/design/unified_error_handling.md | 273 +++++++++ 5 files changed, 1225 insertions(+), 24 deletions(-) create mode 100644 docs/design/blackwell_production_support.md create mode 100644 docs/design/diffusion_ops.md create mode 100644 docs/design/unified_error_handling.md diff --git a/docs/.nav.yml b/docs/.nav.yml index 4ff9493c7..b2ad272e0 100644 --- a/docs/.nav.yml +++ b/docs/.nav.yml @@ -101,30 +101,54 @@ nav: - CI: contributing/ci - Design Documents: - design/index.md - - design/architecture_overview.md + - Architecture Documents: + - Architecture Overview: design/architecture_overview.md - Feature Design: - - design/feature/disaggregated_inference.md - - design/feature/ray_based_execution.md - - design/feature/omni_connectors/ - - design/feature/prefix_caching.md - - design/feature/cfg_parallel.md - - design/feature/expert_parallel.md - - design/feature/sequence_parallel.md - - design/feature/tensor_parallel.md - - design/feature/vae_parallel.md - - design/feature/hsdp.md - - design/feature/cache_dit.md - - design/feature/teacache.md - - design/feature/skip_softmax.md - - design/feature/async_chunk.md - - design/feature/omni_async_output_materialization.md - - design/feature/vae_parallel.md - - design/feature/diffusion_continuous_batching.md + - Runtime and Stage Execution: + - Disaggregated Inference: design/feature/disaggregated_inference.md + - Ray-based Execution: design/feature/ray_based_execution.md + - Async Chunk: design/feature/async_chunk.md + - Async Diffusion Output: design/feature/async_diffusion_output.md + - Async Omni Output Materialization: design/feature/omni_async_output_materialization.md + - Automatic Prefix Caching: design/feature/prefix_caching.md + - Communication: + - OmniConnector Implementations: + - Mooncake Store Connector: design/feature/omni_connectors/mooncake_store_connector.md + - Mooncake Transfer Engine Connector: design/feature/omni_connectors/mooncake_transfer_engine_connector.md + - Mori Transfer Engine Connector: design/feature/omni_connectors/mori_transfer_engine_connector.md + - Shared Memory Connector: design/feature/omni_connectors/shared_memory_connector.md + - Yuanrong Connector: design/feature/omni_connectors/yuanrong_connector.md + - Yuanrong Transfer Engine Connector: design/feature/omni_connectors/yuanrong_transfer_engine_connector.md + - Diffusion Acceleration: + - Parallelism: + - CFG-Parallel: design/feature/cfg_parallel.md + - Expert Parallel: design/feature/expert_parallel.md + - Hybrid Sharded Data Parallel (HSDP): design/feature/hsdp.md + - Pipeline Parallel: design/feature/pipeline_parallel.md + - Sequence Parallel: design/feature/sequence_parallel.md + - Tensor Parallel: design/feature/tensor_parallel.md + - VAE Patch Parallelism: design/feature/vae_parallel.md + - Attention Backends: + - Skip-Softmax: design/feature/skip_softmax.md + - Quantization: user_guide/quantization/overview.md + - Cache-DiT: design/feature/cache_dit.md + - TeaCache: design/feature/teacache.md + - Diffusion Continuous Batching: design/feature/diffusion_continuous_batching.md + - Infrastructure and Performance: + - Prometheus Metrics: design/metrics.md + - Speech Generation Performance Optimizations: design/qwen3_omni_tts_performance_optimization.md - Module Design: - - design/module/ar_module.md - - design/module/dit_module.md - - design/module/entrypoint_module.md - - design/module/async_omni_architecture.md + - Runtime Modules: + - AR Module: design/module/ar_module.md + - DIT Module: design/module/dit_module.md + - Orchestration: + - Entrypoint Module: design/module/entrypoint_module.md + - AsyncOmni Architecture: design/module/async_omni_architecture.md + - RFCs and Experimental Work: + - AR-Diffusion Pipeline Capability: design/ar_diffusion_pipeline_capability.md + - Diffusion Operator Boundary (RFC): design/diffusion_ops.md + - Unified Error Handling (RFC): design/unified_error_handling.md + - Production-Grade NVIDIA Blackwell Support (RFC): design/blackwell_production_support.md - Docs Guide: contributing/DOCS_GUIDE.md - API Reference: - api/README.md diff --git a/docs/design/blackwell_production_support.md b/docs/design/blackwell_production_support.md new file mode 100644 index 000000000..b99e268f8 --- /dev/null +++ b/docs/design/blackwell_production_support.md @@ -0,0 +1,618 @@ +# Production-grade NVIDIA Blackwell support in vLLM-Omni + +> **Status:** Draft for discussion +> +> **Proposed issue title:** `[RFC]: Production-grade NVIDIA Blackwell support in vLLM-Omni` +> +> **vLLM-Omni snapshot:** [`273fc8eb`](https://github.com/vllm-project/vllm-omni/commit/273fc8eb1de084e20a23d03586dec33ac1cbe5ad) +> +> **vLLM baseline:** `v0.26.0` +> +> **Initial scope:** datacenter Blackwell `sm_100` / `sm_103` and +> workstation or consumer Blackwell `sm_120` / `sm_121` + +## Summary + +vLLM-Omni already has substantial Blackwell enablement: + +- the vLLM `v0.26.0` CUDA 13 runtime provides the autoregressive CUDA + foundation; +- [diffusion attention recognizes `sm_100`, `sm_103`, `sm_120`, and + `sm_121`](https://github.com/vllm-project/vllm-omni/blob/273fc8eb1de084e20a23d03586dec33ac1cbe5ad/vllm_omni/platforms/cuda/platform.py#L64-L89); +- datacenter Blackwell can use TRTLLM-generated diffusion attention when its + shape and mask constraints are satisfied; +- cuDNN and FlashInfer provide safe dense diffusion-attention paths for both + Blackwell families; and +- FP8, INT8, ModelOpt, NVFP4, SageAttention 3, and Quack integrations cover + selected models and hardware. + +The remaining problem is not initial bring-up. It is the lack of a precise, +continuously validated support contract. An importable dependency can still +lack a Blackwell kernel, current CUDA CI has no Blackwell resource, and several +quantized paths are documented more broadly than they are exercised on native +Blackwell hardware. + +This RFC proposes: + +1. architecture-aware kernel capability checks with safe, explicit fallbacks; +2. separate validation for datacenter `sm_10x` and workstation `sm_12x`; +3. a small mandatory correctness matrix for attention, sequence parallelism, + representative Omni workloads, and documented quantization paths; +4. a reproducible dependency and release qualification contract; and +5. a distinction between **baseline supported**, **optimized**, and + **experimental** features. + +Optimized-kernel parity between `sm_10x` and `sm_12x` is not required. The two +families expose different hardware features, and a correct cuDNN, FlashInfer, or +SDPA fallback is acceptable for baseline support. + +## Motivation. + +### “Blackwell supported” is currently ambiguous + +The current platform selector contains explicit Blackwell routing and the user +guide publishes +[`sm_120` end-to-end measurements](https://github.com/vllm-project/vllm-omni/blob/273fc8eb1de084e20a23d03586dec33ac1cbe5ad/docs/user_guide/diffusion/attention_backends.md#L170-L181). +This is strong manual evidence that common BF16 diffusion workloads run. + +However, a broad support statement could mean any of the following: + +- the package imports and one model starts; +- common BF16 workloads run on one Blackwell SKU; +- every documented backend fails safely or executes correctly; +- multi-GPU sequence parallelism works; +- documented FP8 or NVFP4 checkpoints use native Blackwell kernels; or +- every Blackwell family has continuous correctness, accuracy, and performance + validation. + +This RFC defines a support claim narrowly enough to be testable. + +### A current failure demonstrates the binary-capability gap + +[`fa3-fwd==0.0.3`](https://github.com/vllm-project/vllm-omni/blob/273fc8eb1de084e20a23d03586dec33ac1cbe5ad/requirements/cuda.txt#L1-L3) +imports on Blackwell, but its extension contains only SM80 and SM90a cubins. +Ring attention treats importability as availability and can therefore launch a +kernel that does not exist for the current device. + +This is reported in +[#5611](https://github.com/vllm-project/vllm-omni/issues/5611). +[#5617](https://github.com/vllm-project/vllm-omni/pull/5617) adds an +architecture-aware SDPA fallback and reports real two-rank validation on +B300/SM103. The defect is narrow, but the lesson is general: importing a Python +module does not prove that an optional extension contains a compatible device +kernel. + +### Current CI cannot continuously validate Blackwell behavior + +CUDA +[test resources currently accept only L4 and H100](https://github.com/vllm-project/vllm-omni/blob/273fc8eb1de084e20a23d03586dec33ac1cbe5ad/tests/helpers/mark.py#L11-L18). +The +[Buildkite hardware presets](https://github.com/vllm-project/vllm-omni/blob/273fc8eb1de084e20a23d03586dec33ac1cbe5ad/.buildkite/common/ci_mirror_hardwares.yml#L20-L190) +likewise have no Blackwell entry on `main`. +[#5543](https://github.com/vllm-project/vllm-omni/pull/5543) is adding B200 +hardware selection, but it is still under review and does not establish an +`sm_120` lane or an architecture-specific pytest contract. + +The +[Qwen3-Omni NVFP4 regression test](https://github.com/vllm-project/vllm-omni/blob/273fc8eb1de084e20a23d03586dec33ac1cbe5ad/tests/e2e/offline_inference/test_qwen3_omni_modelopt_nvfp4_w4a4.py#L14-L18) +explicitly states that its H100 Marlin fallback does not exercise the native +Blackwell FP4 path and leaves a `TODO(B200)` for that coverage. + +### Datacenter and workstation Blackwell are not interchangeable + +`sm_100` / `sm_103` and `sm_120` / `sm_121` share the Blackwell generation but +do not expose identical kernels: + +- TRTLLM-generated diffusion attention is intentionally limited to + datacenter `sm_10x`; +- Quack's `tcgen05` FP8 path is intentionally enabled only on datacenter + `sm_10x`; +- workstation `sm_12x` uses cuDNN or FlashInfer for those workloads; and +- optional FlashInfer, SageAttention, FA4, and low-precision features have + provider-specific architecture matrices. + +Testing B200 alone is therefore insufficient evidence for `sm_120`, while lack +of one optimized `sm_10x` kernel on `sm_120` is not itself a correctness bug. + +## Scope and ownership + +### Initial architecture scope + +| Family | Representative validation target | Initial support intent | +| --- | --- | --- | +| `sm_100` | B200 or GB200 | Continuously validated | +| `sm_103` | B300 or GB300 | Release or periodic compatibility validation | +| `sm_120` | RTX PRO 6000 or equivalent | Continuously validated | +| `sm_121` | Available `sm_121` runner | Compatibility validation when hardware is available | + +This RFC does not silently extend the support claim to `sm_101`, `sm_110`, +future Blackwell variants, Windows, or Jetson deployments. They may reuse the +same capability framework, but each needs an explicit owner and validation +target before being listed as supported. + +### Upstream vLLM versus vLLM-Omni ownership + +| Area | Primary owner | RFC requirement | +| --- | --- | --- | +| Autoregressive LLM attention, CUDA compilation, generic CUDA kernels | vLLM | Consume a Blackwell-capable vLLM release and run a representative Omni smoke test | +| Diffusion attention routing and fallbacks | vLLM-Omni | Validate every default and explicit Blackwell route | +| Ring/Ulysses diffusion sequence parallelism | vLLM-Omni | Provide correct multi-GPU behavior or an actionable early rejection | +| Model-specific audio, vision, talker, tokenizer, and diffusion stages | vLLM-Omni | Validate representative end-to-end paths | +| Generic quantized linear providers | Prefer upstream vLLM | Reuse upstream providers and validate Omni checkpoint/model adaptation | +| Optional provider binaries such as FlashInfer, FA, SageAttention, and Quack | Shared | Declare and test architecture capability before dispatch | + +Generic provider or CUDA defects should be fixed upstream when possible. +vLLM-Omni remains responsible for safe selection, model adaptation, and the +fallback behavior exposed to its users. + +## Goals + +1. Define what vLLM-Omni means when it lists a Blackwell architecture as + supported. +2. Prevent an importable but architecture-incompatible extension from reaching + a CUDA launch. +3. Continuously test representative `sm_100` and `sm_120` behavior. +4. Restore correct ring sequence parallelism on Blackwell, even if the initial + implementation uses SDPA rather than an optimized native kernel. +5. Validate the native Blackwell execution path for every quantization method + documented as supported on Blackwell. +6. Split hardware documentation by `sm_10x` and `sm_12x` where behavior differs. +7. Record enough environment and dispatch information to reproduce a + Blackwell-only failure. + +## Non-goals + +- Requiring identical performance or kernel selection on `sm_10x` and + `sm_12x`. +- Making FA4, quantized diffusion attention, SageAttention 3, or Skip-Softmax a + prerequisite for baseline BF16 support. +- Implementing CUDA-native diffusion MXFP8 or MXFP4 in the first phase. +- Reimplementing generic vLLM, cuDNN, or FlashInfer kernels in vLLM-Omni. +- Validating every model and every quantization combination on every + Blackwell SKU. +- Treating manual benchmark results as a replacement for recurring CI. +- Expanding the initial claim to unlisted Blackwell variants without hardware + validation. + +## Support contract + +### Support levels + +**Baseline supported** + +- the documented installation resolves to a compatible vLLM, PyTorch, CUDA, + cuDNN, and provider set; +- the runtime selects only a kernel known to support the current architecture; +- an unsupported explicit backend fails during configuration or warmup with an + actionable error; +- a documented safe fallback produces correct output; +- representative eager and compiled paths pass recurring hardware CI; and +- applicable multi-GPU and documented quantization smokes pass. + +**Optimized** + +- the architecture uses an intentional accelerated provider rather than the + conservative fallback; +- accuracy is compared with the architecture-local BF16 baseline; and +- performance is monitored against a per-SKU baseline. + +**Experimental** + +- the feature is opt-in; +- its architecture or model matrix may be incomplete; and +- it is not part of the release-blocking baseline. + +### Proposed compatibility matrix + +| Capability | `sm_100` / `sm_103` | `sm_120` / `sm_121` | Required RFC exit state | +| --- | --- | --- | --- | +| AR/LLM CUDA path | Inherited from vLLM; not directly covered by current Blackwell CI | Same | Representative Omni AR-stage smoke on each continuously validated family | +| Dense BF16 diffusion attention | TRTLLM when compatible; otherwise cuDNN/FlashInfer/SDPA | cuDNN/FlashInfer/SDPA | Default route and fallback tested on real hardware | +| Ring sequence parallelism | B200 failure reported in #5611 | Same import-only selection risk is unverified | Correct SDPA fallback plus real two-rank Blackwell test | +| Ulysses sequence parallelism | Expected to use the selected local backend | Same | Existing multi-GPU smoke rerun on B200; failures must identify unsupported shapes | +| TRTLLM diffusion attention | Supported only on datacenter `sm_10x` with provider, shape, and mask constraints | Unsupported by design | Positive `sm_10x` test and early `sm_12x` rejection test | +| FP8 and INT8 linear paths | Documented support; native path not broadly covered | Documented support; provider differs | One representative diffusion checkpoint per family | +| ModelOpt NVFP4 | Documented Blackwell path; native regression coverage incomplete | Broad Blackwell claim is not separately validated on `sm_12x` | Native `sm_10x` smoke; validate `sm_12x` separately or narrow its documentation | +| Diffusion-native MXFP8 / MXFP4 | Unsupported on CUDA today | Unsupported on CUDA today | Documentation says unsupported; configuration fails early | +| Quack FP8 | Optimized on datacenter `sm_10x` | Unsupported by hardware; FlashInfer fallback | Positive `sm_10x` and negative `sm_12x` selection tests | +| SageAttention 3, FA4, and quantized attention | Experimental or in flight | Experimental or provider-dependent | Not release-blocking until promoted through a follow-up RFC or support update | + +## Proposed Change. + +### 1. Centralize architecture and provider capability checks + +Introduce one internal CUDA architecture classification used by platform +routing, attention backends, ring attention, quantization providers, and tests. +It should distinguish at least: + +- datacenter Blackwell `sm_10x`; +- workstation or consumer Blackwell `sm_12x`; +- explicitly recognized minor architectures; and +- unknown or unvalidated variants. + +Every optional provider selected by vLLM-Omni should expose or be wrapped by a +capability predicate that answers whether it can execute the requested +operation on the actual device. The predicate may use: + +- provider-published architecture metadata; +- a known contract for an exactly pinned provider version; +- an upstream vLLM provider capability API; or +- a bounded startup probe when metadata is unavailable and probing cannot + terminate the process. + +Importability alone is insufficient. + +Selection must have one of three observable outcomes: + +1. select a compatible provider; +2. select a documented safe fallback and log the reason once; or +3. reject an explicitly requested unsupported provider before the first real + request. + +The runtime log should identify the device capability, selected provider, and +fallback reason. It must not claim that a fallback provider is the optimized +path. + +The first implementation should: + +- land the architecture-aware ring fallback from #5617; +- audit `vllm_omni.diffusion.attention.backends.utils.fa`; +- audit direct `flash_attn`, `fa3_fwd_interface`, and + `flash_attn_interface` imports in model-specific code; and +- make the FA3 Hub and explicit FlashAttention gates match their provider + contracts. + +### 2. Add architecture-specific CI resources + +Extend the CUDA resource helpers and Buildkite presets with architecture-honest +names. A test running on B200 must not continue to advertise itself as an H100 +test merely to reuse an existing marker. + +Recommended representative resources: + +- `B200`, with one-, two-, and four-GPU presets as infrastructure permits; and +- `RTXPRO6000` or another stable `sm_120` product name, initially single-GPU. + +The `MIRROR_HW` mechanism proposed by #5543 can choose the physical preset, but +pytest selection should still express the architecture actually required by a +test. + +#### CI levels + +| Level | Hardware | Required coverage | Blocking policy | +| --- | --- | --- | --- | +| L1/L2 | CPU plus existing CUDA | Architecture classifier, provider-capability, routing, early-error, and fallback unit tests | Required on affected PRs | +| L3 | One or two B200 GPUs, change-selected | Startup/kernel smoke; two-rank ring when relevant; native quantization smoke when relevant | Required pre-merge for Blackwell-sensitive PRs through the existing `merge-test` trigger; main-branch runs provide detection and recovery | +| L4 | B200 and `sm_120` representative | End-to-end BF16, compiled path, two-rank B200 ring, representative Omni AR stage, and documented FP8/NVFP4 paths—or an explicitly narrowed support status where qualification is incomplete | Nightly release signal | +| L5/release | B300/SM103 and other available variants | Compatibility smoke, extended models, accuracy, and performance | Required before expanding or retaining the published variant matrix | + +Infrastructure failures should be reported separately from product failures. +An architecture should not be promoted to baseline supported until its required +jobs complete seven consecutive scheduled runs without a product failure. + +Blackwell-sensitive source paths should automatically request, or be required +by project policy to use, the existing `merge-test` trigger. A label that can be +silently omitted is not a merge gate. Required Blackwell jobs must execute with +zero unexpected skips or xfails; a green job in which the architecture-specific +test never ran is not evidence of support. + +### 3. Use a small representative model matrix + +The recurring matrix should cover behavior classes rather than every model: + +- one mask-free video DiT that can exercise the datacenter TRTLLM route, such + as Wan2.2; +- one mask-using image or video DiT that exercises the cuDNN path, such as + Qwen-Image or HunyuanVideo-1.5; +- one multi-stage or Omni workload that reaches an upstream vLLM AR stage and + at least one Omni-owned non-AR component; +- one audio or TTS integration canary that exercises a model-specific encoder + or attention adapter; +- one native Blackwell NVFP4 checkpoint smoke. + +Tiny or random-weight models should cover routing, shapes, compilation, and +collective behavior in lower levels. Full checkpoints should be reserved for +nightly accuracy and integration validation. + +Large Omni checkpoints do not need to fit on a single workstation GPU to call +the `sm_120` family supported. The `sm_120` lane must instead validate +representative components and at least one end-to-end workload that fits the +target resource, while documenting memory-driven model exclusions. + +### 4. Make quantization claims path-specific + +Quantization documentation and tests should distinguish: + +- generic upstream vLLM FP8, INT8, and NVFP4 providers; +- ModelOpt checkpoint adaptation in vLLM-Omni; +- online versus pre-quantized checkpoints; +- diffusion versus AR/thinker stages; and +- datacenter versus workstation provider selection. + +Each Blackwell-supported quantization row needs: + +1. proof that the intended quantized provider is selected; +2. evidence that weights remain in the intended representation or reduce + memory as expected; +3. a finite and non-degenerate output assertion; and +4. an accuracy comparison with a same-architecture BF16 baseline. + +The open report in +[#4202](https://github.com/vllm-project/vllm-omni/issues/4202) should be +reproduced or closed against the RFC baseline. Because that report used +vLLM-Omni `0.20.0`, it is evidence requiring current reproduction, not proof +that `0.26.0` is still broken. + +CUDA-native diffusion MXFP8 and MXFP4 should be marked unsupported until a +separate implementation and validation effort lands. “Not verified” is too +ambiguous when +[MXFP8](https://github.com/vllm-project/vllm-omni/blob/273fc8eb1de084e20a23d03586dec33ac1cbe5ad/vllm_omni/quantization/mxfp8_config.py#L120-L147) +and +[MXFP4](https://github.com/vllm-project/vllm-omni/blob/273fc8eb1de084e20a23d03586dec33ac1cbe5ad/vllm_omni/quantization/mxfp4_config.py#L142-L168) +explicitly raise `NotImplementedError` on CUDA. + +### 5. Qualify the dependency and release environment + +vLLM-Omni owns no in-tree CUDA or C++ extension for these paths; its Blackwell +binary support is supplied by vLLM, PyTorch, cuDNN, FlashInfer, FA packages, +SageAttention, Quack, and other optional providers. + +Release qualification should therefore record: + +- GPU product and compute capability; +- driver, CUDA runtime, PyTorch, and cuDNN versions; +- vLLM and vLLM-Omni versions; +- optional provider package versions; +- provider-reported or known compiled architectures; and +- the selected attention and quantized-linear providers. + +The supported installation path must resolve to a tested combination. Possible +implementations include an exact vLLM dependency, a generated compatibility +constraint, or a startup error for an unsupported major/minor combination. +The current warning-only behavior is insufficient for an ABI-critical mismatch. + +CUDA 13 driver requirements and the tested container tag should be stated in +the installation guide. Published containers should use an immutable base +reference or otherwise record the resolved base-image digest in release +artifacts. + +### 6. Publish an architecture-specific support table + +The user guide should avoid a single “Blackwell SM100+” row when provider +selection differs by family. It should list: + +- supported architecture families and representative tested GPUs; +- the tested vLLM/CUDA/driver baseline; +- default and fallback attention providers; +- explicit unsupported combinations; +- quantization status by model stage and provider; and +- whether each item is baseline supported, optimized, experimental, + unsupported, or unverified. + +Release notes should link the exact CI or validation artifact used to retain a +support claim. + +## Implementation phases and exit criteria + +### Phase 0: Correctness floor + +- [ ] Merge an architecture-aware ring fallback for #5611. +- [ ] Add unit tests for `sm_100`, `sm_103`, `sm_120`, `sm_121`, and unknown + architectures. +- [ ] Audit import-only FA capability checks. +- [ ] Make unsupported explicit backends fail before request execution. +- [ ] Mark CUDA diffusion MXFP8/MXFP4 unsupported in code-facing documentation. + +**Exit criterion:** every known Blackwell route either selects a compatible +provider, takes a tested safe fallback, or fails early with an actionable +message. + +### Phase 1: Datacenter Blackwell baseline + +- [ ] Land B200 Buildkite presets and honest pytest markers. +- [ ] Add real one-GPU default-routing and two-GPU ring tests. +- [ ] Add BF16 eager and compiled diffusion smokes. +- [ ] Add one representative Omni AR-stage smoke. +- [ ] Add native FP8 and NVFP4 smokes with provider and output assertions. +- [ ] Establish per-B200 accuracy baselines rather than reusing H100 thresholds. + +**Exit criterion:** required B200 jobs pass seven consecutive scheduled runs +without a product failure or unexpected architecture-specific skip. + +### Phase 2: Workstation Blackwell baseline + +- [ ] Add an `sm_120` hardware preset and marker. +- [ ] Validate the cuDNN default and FlashInfer/SDPA fallbacks. +- [ ] Assert that TRTLLM and Quack are not auto-selected. +- [ ] Run a fitting end-to-end workload in eager and compiled modes. +- [ ] Validate representative FP8 and NVFP4 paths or narrow their documented + support. +- [ ] Establish per-`sm_120` accuracy baselines. + +**Exit criterion:** required `sm_120` jobs pass seven consecutive scheduled +runs without a product failure or unexpected architecture-specific skip. + +### Phase 3: Optimization and expanded coverage + +- [ ] Add stable per-SKU performance monitoring. +- [ ] Evaluate FA4 without making it a baseline requirement. +- [ ] Evaluate Blackwell quantized diffusion attention and typed SAGE paths. +- [ ] Add B300/SM103 periodic or release validation. +- [ ] Add further models and `sm_121` when hardware and owners are available. + +**Exit criterion:** individual optimized features are promoted only after they +have architecture-specific correctness, accuracy, and performance evidence. + +## Correctness and testing plan + +### Routing and binary compatibility + +Table-driven tests should cover: + +- every recognized Blackwell architecture family; +- cuDNN above and below the supported threshold; +- FlashInfer present, absent, partially installed, and missing the required + symbol; +- an FA Python module that imports while its version contract excludes the + current architecture; +- explicit versus automatic backend selection; +- mask-free versus mask-using models; +- supported and unsupported head dimensions; and +- datacenter-only providers on `sm_12x`. + +### Multi-GPU + +At least one real two-rank NCCL test must exercise: + +- ring fallback selection; +- finite BF16 outputs; +- output shape and shard ownership; +- comparison with a single-GPU or non-ring reference within a documented + tolerance; and +- eager and compiled behavior where supported. + +Ulysses and hybrid Ulysses/Ring should retain their existing shape constraints +and fail clearly when an input is unsupported. + +### Quantization + +Quantization tests should assert provider selection and representation, not +only that a request returns successfully. Accuracy comparisons should use the +same architecture, prompt, seed, dimensions, and scheduler settings as the BF16 +reference. + +Architecture-specific thresholds are allowed. They must be derived from +measured numerical variation and may not be loosened merely to make a new +runner pass. + +### Performance + +Performance jobs should initially run in monitor-only mode until enough samples +exist to estimate variance. A later gate can use a threshold based on both a +fixed practical floor and observed noise, for example the larger of 10% or +three standard deviations. + +Performance regression on an optimized provider is not automatically a +baseline-correctness failure if the documented safe fallback remains correct. +It blocks promotion or retention of the **optimized** label. + +## Compatibility and rollout + +No public API change is required for the initial phases. Users may observe: + +- a previously crashing backend falling back to SDPA; +- an explicit unsupported backend failing earlier; +- more precise startup logs; +- CUDA MXFP8/MXFP4 documentation changing from unverified to unsupported; and +- narrower quantization claims for combinations without native hardware + evidence. + +These are intentional correctness changes. Any fallback that can materially +change latency should emit one clear warning and be included in release notes. + +CI rollout should be additive. Existing H100 and L4 coverage remains necessary; +Blackwell jobs validate architecture-specific behavior and must not replace all +pre-Blackwell regression coverage. + +## Dependencies and risks + +| Risk | Mitigation | +| --- | --- | +| Scarce or unstable Blackwell runners | Keep lower-level routing tests hardware-free; separate infrastructure failures; use targeted L3 and recurring L4/L5 jobs | +| Third-party wheels change embedded cubins or ABI | Pin or record exact versions, declare architecture contracts, and verify provider selection at startup | +| B200 results are incorrectly generalized to `sm_120` | Maintain separate family rows, runners, and accuracy baselines | +| Full-model CI becomes too expensive | Use tiny models for routing/kernel tests and a minimal representative full-model matrix nightly | +| Cross-GPU numerical drift creates false accuracy failures | Use architecture-local baselines and thresholds derived from repeated measurements | +| Omni duplicates upstream vLLM work | Fix generic provider contracts upstream and keep only routing/model adaptation in Omni | +| Safe fallback hides a performance regression | Log selected providers, distinguish baseline from optimized support, and monitor per-SKU performance | + +## Alternatives considered + +### Rely only on upstream vLLM Blackwell support + +Rejected. Upstream owns the AR CUDA foundation, but vLLM-Omni owns diffusion +attention routing, ring parallelism, model-specific stages, and checkpoint +adaptation. The current ring defect exists above the upstream boundary. + +### Validate only B200 and call all Blackwell supported + +Rejected. Datacenter `sm_10x` and workstation `sm_12x` intentionally use +different providers and tensor-core features. + +### Wait for FA4 before declaring baseline support + +Rejected. cuDNN, FlashInfer, and SDPA already provide viable correctness paths. +FA4 is an optimization and should be promoted independently. + +### Require full kernel and model parity before any support claim + +Rejected. It would make support unattainable and would ignore legitimate +hardware differences. A narrow tested baseline plus explicit experimental +features is clearer. + +### Keep using manual benchmark evidence + +Rejected as the only mechanism. Manual results are useful for bring-up and +optimization decisions but do not prevent regressions in routing, dependencies, +or optional binary packages. + +## Open questions + +1. Should `sm_103` be continuously tested, or is a B300 release qualification + sufficient while B200 remains the recurring `sm_10x` representative? +2. Which stable `sm_120` product and runner pool can the project commit to? +3. Which mask-using diffusion and fitting Omni workloads should form the + smallest representative matrix? +4. Should vLLM-Omni encode an exact vLLM dependency, a generated compatibility + range, or a hard startup check? +5. Should provider capability metadata become an upstream vLLM interface? +6. Is a correct SDPA ring fallback sufficient for baseline support, with a + native Blackwell ring kernel tracked separately? +7. Should native CUDA diffusion MXFP8/MXFP4 remain out of scope, or should their + current configuration names be rejected earlier at CLI/config parsing? +8. What stability window and performance threshold should be required before + promoting an experimental provider to optimized? + +## Feedback Period. + +Feedback is requested for two weeks after publication. + +## CC List. + +TBD: CUDA platform, diffusion attention, sequence parallelism, quantization, +CI infrastructure, and upstream vLLM kernel maintainers. + +## Any Other Things. + +### Suggested contribution split + +The work can land as small, independently reviewable changes: + +1. ring fallback and architecture-contract tests; +2. shared CUDA architecture classification; +3. FA/provider capability audit; +4. B200 resource markers and CI presets; +5. B200 routing, ring, and quantization smokes; +6. `sm_120` resource and CI lane; +7. architecture-specific documentation and release manifest; and +8. optional performance-provider follow-ups. + +### References + +- [vLLM GPU installation and Blackwell CUDA requirement](https://docs.vllm.ai/en/latest/getting_started/installation/gpu/) +- [#3079: CUDNN_ATTN and FLASHINFER_ATTN Blackwell auto-routing](https://github.com/vllm-project/vllm-omni/pull/3079) +- [#3015: SageAttention 3 Blackwell diffusion backend](https://github.com/vllm-project/vllm-omni/pull/3015) +- [#4025: Qwen3-Omni NVFP4 W4A4 serving on Blackwell](https://github.com/vllm-project/vllm-omni/pull/4025) +- [#4241: Quack FP8 on datacenter Blackwell](https://github.com/vllm-project/vllm-omni/pull/4241) +- [#4817: Quack gating for `sm_10x` versus `sm_12x`](https://github.com/vllm-project/vllm-omni/pull/4817) +- [#5288: TRTLLM diffusion-attention roadmap](https://github.com/vllm-project/vllm-omni/issues/5288) +- [#5283: TRTLLM diffusion attention with Skip-Softmax](https://github.com/vllm-project/vllm-omni/pull/5283) +- [#5611: B200 ring-attention failure](https://github.com/vllm-project/vllm-omni/issues/5611) +- [#5617: architecture-aware ring fallback](https://github.com/vllm-project/vllm-omni/pull/5617) +- [#5543: B200 Buildkite hardware selection](https://github.com/vllm-project/vllm-omni/pull/5543) +- [#4202: ModelOpt FP8 checkpoint behavior on Blackwell](https://github.com/vllm-project/vllm-omni/issues/4202) +- [#1959: NVFP4 quantization support for diffusion models](https://github.com/vllm-project/vllm-omni/issues/1959) +- [#4858: FA4 diffusion attention](https://github.com/vllm-project/vllm-omni/pull/4858) +- [#5344: Blackwell QK16/V8 FlashInfer attention](https://github.com/vllm-project/vllm-omni/pull/5344) +- [#5509: typed SAGE for TRTLLM diffusion attention](https://github.com/vllm-project/vllm-omni/pull/5509) +- [#3340: consolidate Omni/TTS FlashAttention dispatch through vLLM](https://github.com/vllm-project/vllm-omni/pull/3340) diff --git a/docs/design/diffusion_ops.md b/docs/design/diffusion_ops.md new file mode 100644 index 000000000..2c721ccfc --- /dev/null +++ b/docs/design/diffusion_ops.md @@ -0,0 +1,231 @@ + + +# RFC: Establish an operator boundary for vLLM-Omni diffusion + +> **Status:** Draft for discussion +> +> **vLLM-Omni snapshot:** [`92534733`](https://github.com/vllm-project/vllm-omni/commit/925347332f8190d172dfa8ac0c307a5cfa2320cb) +> +> **Upstream vLLM snapshot:** [`fa2a2589`](https://github.com/vllm-project/vllm/commit/fa2a2589bd2a1fce0851df7fd42ffb54b6195f04) + +## Summary + +vLLM has an operator boundary, most visibly through `_custom_ops.py` and its +kernel-provider interfaces. Model code can use optimized kernels without +depending on native extension names, packed layouts, allocation details, or +platform-specific dispatch. + +vLLM-Omni already reuses much of that machinery. However, diffusion support has +also accumulated local Triton kernels, `torch.library` custom ops, platform +calls, third-party providers, and compatibility patches. These cover semantics +that are sometimes different from autoregressive inference: + +- N-dimensional and non-contiguous activations; +- routed mixture-of-transformers (MoT) computation; +- diffusion-specific attention and quantized attention; +- platform-specific microscaling formats; and +- stateful diffusion cache operations. + +This RFC asks whether vLLM-Omni should define an explicit diffusion operator +boundary. “Independent” means independently defined semantics, not independent +reimplementation: generic kernels should continue to come from upstream vLLM. + +Approval would only establish ownership principles. It would not approve a +refactor, public API, file layout, or migration plan. + +## Motivation + +### Diffusion needs a stronger tensor contract + +Diffusion linears commonly receive tensors shaped `(..., K)` and must return +`(..., N)`, preserving all leading dimensions. Inputs may also be +non-contiguous. + +The current code adapts upstream FP8 paths by: + +- making ScaledMM activations contiguous before a two-dimensional `view`; and +- restoring leading dimensions when the FlashInfer provider returns a + two-dimensional result. + +Several quantization methods independently perform similar flatten/restore +logic. These are provider-contract requirements, not new GEMM algorithms. + +### Some operations are genuinely diffusion-specific + +Examples include: + +- MoT kernels that route text and VAE tokens to different weights inside one + fused GEMM; +- diffusion attention with its own layouts, masks, parallel strategies, and + optional quantization; +- NPU FP8 attention that rotates and block-quantizes Q/K/V on each denoising + step; and +- experimental paged attention that writes into a process-owned diffusion KV + pool. + +These semantics do not map cleanly to generic dense-linear or autoregressive +attention interfaces. + +### Quantization mixes policy, checkpoint handling, and kernels + +Current quantization code combines: + +1. deciding which stages and layers to quantize; +2. loading, remapping, and packing weights and scales; and +3. selecting and invoking a hardware kernel. + +The first two are often Omni-specific. The third is frequently generic and +already supported upstream. Without a clear boundary, it is difficult to tell +when Omni is extending vLLM and when it is duplicating it. + +## Current status + +The audited snapshot has no project-owned C++, CUDA, or HIP extension sources. +Its optimized operator surface is Python- and provider-driven: + +| Mechanism | Current status | +| --- | --- | +| Upstream vLLM ops/providers | Default path for generic quantization, GEMM, normalization, and related primitives | +| Diffusion `CustomOp` | Construction-time platform dispatch for norm, RoPE, AdaLayerNorm, and MoT RMSNorm | +| Attention registry | Nine default diffusion attention backends with platform overrides | +| Local Triton | Seven decorated functions, primarily MoT plus one model-specific fusion | +| `torch.library` | Three custom ops: SageAttention 3, ROCm MXFP4, and experimental paged write-attention | +| External providers | Direct use of `torch_npu`, AITER, FlashInfer, SageAttention, BitsAndBytes, and Quack | +| Compatibility patches | FP8 contiguity/output shape, NVFP4 scale sanitization, Quack selection, and scoped provider forcing | + +### Quantization snapshot + +| Path | Kernel source | Omni-specific behavior | +| --- | --- | --- | +| FP8 / ModelOpt / NVFP4 | Upstream vLLM | Checkpoint adaptation, diffusion shape fixes, HSDP weight views, scale sanitization | +| CUDA INT8 | Upstream vLLM | Configuration and online/offline weight handling | +| NPU INT8 and MXFP8 | `torch_npu` | Layouts, scales, bias/dtype handling, and shape restoration | +| XPU MXFP8 | Upstream vLLM | Local N-dimensional wrapper | +| NPU MXFP4 | `torch_npu` | Single-scale and diffusion-oriented DualScale formats | +| ROCm MXFP4 | Local AITER custom op | Overlaps with an AITER provider now present upstream | +| BitsAndBytes W4 | BitsAndBytes | Load-time quantization and shape adaptation | +| Quack FP8 | Quack CuteDSL | Installed by replacing the selected FlashInfer method | +| MoT FP8/INT8 | Local Triton plus upstream FP8 activation quantization | Routed text/VAE semantics | + +Upstream now has provider registries for FP8, INT8, NVFP4, MXFP8, and MXFP4 +across supported platforms, plus `register_linear_kernel()`. This should remain +the preferred home for model-agnostic linear kernels. + +### Problems exposed by the current structure + +- Tensor shape, contiguity, dtype, bias, packing, and scale contracts are not + expressed uniformly across providers. +- Global patches solve real problems but make provider ownership, ordering, and + version compatibility difficult to inspect. +- FakeTensor, `torch.compile`, fallback, mutation, and aliasing behavior varies + by integration. +- Platform capability information is duplicated across code and documentation + and has already drifted. +- Tests cover configuration and weight layouts well, but quantized MoT, Quack + execution, real ROCm MXFP4, and cross-provider compile behavior remain thin. + +## Direction proposed for discussion + +```text +diffusion model or layer + | + v +diffusion semantic contract, only where upstream is insufficient + | + +----> upstream vLLM provider + +----> Omni-owned diffusion provider + +----> platform or out-of-tree provider +``` + +Suggested principles: + +1. **Reuse upstream by default.** Do not wrap or fork a generic kernel whose + upstream contract already works for diffusion. +2. **Fix generic contracts upstream.** Shape preservation, non-contiguous input + handling, generic packing, and provider registration should live in vLLM + when they benefit more than diffusion. +3. **Own diffusion semantics in Omni.** MoT routing, quantized diffusion + attention, and diffusion cache behavior are reasonable Omni responsibilities. +4. **Separate policy from execution.** Stage selection, ignored layers, + checkpoint adaptation, and quality-sensitive BF16 fallback are not kernel + responsibilities. +5. **Treat hardware integrations as providers.** Model code should not depend + directly on vendor ABI, packed layouts, or optional package internals. +6. **Make compilation and state explicit.** Opaque kernels need deliberate fake + behavior, mutation/aliasing semantics, and fallback rules. +7. **Prefer scoped selection over global replacement.** Active providers and + fallbacks should be inspectable and limited to the intended operation or + stage. +8. **Do not require one monolithic file.** The boundary could be a small package + or extensions to existing registries; copying upstream `_custom_ops.py` is + not the goal. + +Likely upstream-owned: + +- generic dense-linear quantization kernels and activation quantization; +- provider-independent shape/dtype contracts; and +- generic provider registration and selection. + +Likely Omni-owned: + +- routed MoT operators; +- dynamically quantized diffusion attention; +- diffusion-specific cache operations; and +- formats or semantics not represented upstream, such as the current NPU + MXFP4 DualScale path. + +## Non-goals + +- Defining a detailed refactor or migration sequence. +- Choosing final module names or public Python signatures. +- Reimplementing vLLM's generic custom ops or linear kernels. +- Adding an Omni native extension loader without a demonstrated need. +- Changing supported hardware or claiming unvalidated quantization support. +- Removing compatibility patches before replacement behavior is validated. + +## Open questions + +1. Is an explicit diffusion operator boundary needed, or are the current layer + and backend interfaces sufficient? +2. Should the boundary remain diffusion-only or eventually cover other + non-autoregressive Omni stages? +3. What minimum shape, layout, dtype, bias, and compilation contract should + every provider satisfy? +4. Which current adaptations are generic upstream fixes versus + diffusion-specific behavior? +5. Should quantized diffusion attention be an attention backend, an operator + provider, or both? +6. Should platform providers live in Omni core, platform projects, or + out-of-tree packages? +7. Is upstream `register_linear_kernel()` sufficient for provider priority, + diagnostics, explicit selection, and fallback? +8. How should stateful operators declare process-global mutation to eager mode + and `torch.compile`? +9. What validation should be required before an experimental provider becomes + a supported default? + +## Feedback requested + +Feedback is requested from upstream kernel, diffusion, quantization, platform, +attention, and `torch.compile` maintainers. Particularly useful comments would +identify: + +- an operation that clearly belongs upstream or clearly belongs in Omni; +- a missing provider contract; +- a compatibility patch that should remain local or move upstream; or +- a platform integration that needs an out-of-tree extension point. + +If there is agreement on the ownership principles, a follow-up design can +compare concrete API options. If not, this RFC should document why the existing +abstractions are sufficient. + +## References + +- [vLLM `_custom_ops.py`](https://github.com/vllm-project/vllm/blob/fa2a2589bd2a1fce0851df7fd42ffb54b6195f04/vllm/_custom_ops.py) +- [vLLM linear-kernel registry](https://github.com/vllm-project/vllm/blob/fa2a2589bd2a1fce0851df7fd42ffb54b6195f04/vllm/model_executor/kernels/linear/__init__.py) +- [vLLM-Omni quantization factory](https://github.com/vllm-project/vllm-omni/blob/925347332f8190d172dfa8ac0c307a5cfa2320cb/vllm_omni/quantization/factory.py) +- [Diffusion attention registry](https://github.com/vllm-project/vllm-omni/blob/925347332f8190d172dfa8ac0c307a5cfa2320cb/vllm_omni/diffusion/attention/backends/registry.py) +- [Diffusion `CustomOp`](https://github.com/vllm-project/vllm-omni/blob/925347332f8190d172dfa8ac0c307a5cfa2320cb/vllm_omni/diffusion/layers/custom_op.py) +- [Compatibility patches](https://github.com/vllm-project/vllm-omni/blob/925347332f8190d172dfa8ac0c307a5cfa2320cb/vllm_omni/patch.py) +- [MoT GEMM](https://github.com/vllm-project/vllm-omni/blob/925347332f8190d172dfa8ac0c307a5cfa2320cb/vllm_omni/diffusion/layers/mot/ops/mot_gemm.py) +- [NPU FP8 diffusion attention](https://github.com/vllm-project/vllm-omni/blob/925347332f8190d172dfa8ac0c307a5cfa2320cb/vllm_omni/platforms/npu/quant/kv_quant_npu.py) diff --git a/docs/design/index.md b/docs/design/index.md index 0da17cce4..67401dffe 100644 --- a/docs/design/index.md +++ b/docs/design/index.md @@ -1,6 +1,8 @@ # Design Documents -This section contains design documents and architecture specifications for vLLM-Omni. +This section contains design documents and architecture specifications for +vLLM-Omni. The sidebar groups documents by the system concern they describe, +while the source paths remain stable so existing links continue to work. ## Architecture Documents @@ -8,17 +10,70 @@ This section contains design documents and architecture specifications for vLLM- ## Feature Design Documents +### Runtime and stage execution + - [Disaggregated Inference](feature/disaggregated_inference.md) - [Ray-based Execution](feature/ray_based_execution.md) +- [Async Chunk](feature/async_chunk.md) +- [Async Diffusion Output](feature/async_diffusion_output.md) - [Async Omni Output Materialization](feature/omni_async_output_materialization.md) +- [Automatic Prefix Caching in Omni Models](feature/prefix_caching.md) + +### Communication + +#### OmniConnector implementations + +- [Mooncake Store Connector](feature/omni_connectors/mooncake_store_connector.md) +- [Mooncake Transfer Engine Connector](feature/omni_connectors/mooncake_transfer_engine_connector.md) +- [Mori Transfer Engine Connector](feature/omni_connectors/mori_transfer_engine_connector.md) +- [Shared Memory Connector](feature/omni_connectors/shared_memory_connector.md) +- [Yuanrong Connector](feature/omni_connectors/yuanrong_connector.md) +- [Yuanrong Transfer Engine Connector](feature/omni_connectors/yuanrong_transfer_engine_connector.md) + +### Diffusion acceleration + +#### Parallelism + +- [CFG-Parallel](feature/cfg_parallel.md) +- [Expert Parallel](feature/expert_parallel.md) +- [Hybrid Sharded Data Parallel (HSDP)](feature/hsdp.md) +- [Pipeline Parallel](feature/pipeline_parallel.md) +- [Sequence Parallel](feature/sequence_parallel.md) +- [Tensor Parallel](feature/tensor_parallel.md) +- [VAE Patch Parallelism](feature/vae_parallel.md) + +#### Attention Backends + +- [Skip-Softmax](feature/skip_softmax.md) + +#### Quantization + +- [Quantization Overview](../user_guide/quantization/overview.md) + +- [Cache-DiT](feature/cache_dit.md) +- [TeaCache](feature/teacache.md) - [Diffusion Continuous Batching](feature/diffusion_continuous_batching.md) -## Infrastructure Design Documents +## Infrastructure and Performance - [Prometheus Metrics](metrics.md) +- [Speech Generation Performance Optimizations](qwen3_omni_tts_performance_optimization.md) ## Module Design Documents +### Runtime Modules + - [AR Module](module/ar_module.md) - [DIT Module](module/dit_module.md) + +### Orchestration + - [Entrypoint Module](module/entrypoint_module.md) +- [AsyncOmni Architecture (Qwen3-Omni Example)](module/async_omni_architecture.md) + +## RFCs and Experimental Work + +- [AR-Diffusion Pipeline Capability](ar_diffusion_pipeline_capability.md) +- [Diffusion Operator Boundary (RFC)](diffusion_ops.md) +- [Unified Error Handling (RFC)](unified_error_handling.md) +- [Production-Grade NVIDIA Blackwell Support (RFC)](blackwell_production_support.md) diff --git a/docs/design/unified_error_handling.md b/docs/design/unified_error_handling.md new file mode 100644 index 000000000..8a8a87b4c --- /dev/null +++ b/docs/design/unified_error_handling.md @@ -0,0 +1,273 @@ + + +# RFC: Align vLLM-Omni error handling with vLLM + +> **Status:** Draft for discussion +> **Feedback period:** Two weeks after publication +> **vLLM-Omni snapshot:** [`1b718f7b`](https://github.com/vllm-project/vllm-omni/commit/1b718f7b1529bf7b59450de2dba0f820ef676dc8) +> **Upstream vLLM snapshot:** [`fa2a2589`](https://github.com/vllm-project/vllm/commit/fa2a2589bd2a1fce0851df7fd42ffb54b6195f04) +> **Upstream direction:** [vLLM RFC #48227](https://github.com/vllm-project/vllm/issues/48227) and [vLLM PR #49665](https://github.com/vllm-project/vllm/pull/49665) + +## Motivation + +The error fragmentation described in upstream vLLM also exists in vLLM-Omni +and is amplified by multi-stage execution and additional serving transports. +In the audited snapshot, `vllm_omni/` contains: + +- 2,147 `raise ValueError` sites, including 171 in `entrypoints/`; +- 166 `raise TypeError` sites; +- 460 `raise RuntimeError` sites; +- no direct `raise VLLMValidationError` site; and +- one direct `raise GuardrailViolationError` site. + +These counts do not imply that every raw exception should become public. They +show that raw Python types currently represent request validation, internal +invariants, configuration errors, model failures, and control flow, while +semantic classification is rare. + +This causes four concrete problems: + +1. **Internal failures can be reported as HTTP 400.** Several entrypoints treat + every `ValueError` as invalid user input: + + ```python + except OmniClientError as exc: + raise HTTPException(exc.status_code, exc.message) from exc + except ValueError as exc: + raise HTTPException(400, str(exc)) from exc + except Exception as exc: + raise HTTPException(500, str(exc)) from exc + ``` + + A `ValueError` from parsing, model execution, output conversion, or + configuration can therefore be reported as `400 Bad Request` instead of + `500 Internal Server Error`. The safer default is that only an explicit + client-error type becomes 4xx. + +2. **Ownership is guessed after crossing stage boundaries.** Current receivers + reconstruct a client error when the serialized status is in the 4xx range + and otherwise construct an engine error. This derives semantic ownership + from a status chosen earlier, so a wrong classification cannot be recovered. + Parameter information and nonfatal stage context are also not consistently + preserved. + +3. **The same failure has different public contracts.** HTTP routes, SSE, + WebSocket, speech helpers, and asynchronous video jobs convert failures + independently. [vLLM-Omni PR #3969](https://github.com/vllm-project/vllm-omni/pull/3969) + already fixed one result of this fragmentation: voice endpoints returning + an error object with HTTP 200. + +4. **Prometheus status metrics are fragile.** + [vLLM PR #44303](https://github.com/vllm-project/vllm/pull/44303) showed that + a client can receive 4xx while `http_requests_total` records 5xx when an + exception crosses the Prometheus middleware before reaching an outer + catch-all handler. Omni adds its own exceptions and handlers on top of the + upstream application, so relying on per-type registration can reproduce the + mismatch. Returning an error model with default HTTP 200 creates the inverse + problem: failed requests are recorded as 2xx. + +## Proposed change + +Adopt upstream vLLM's semantic exception hierarchy for in-process errors and +add one Omni-owned serializable record for stage/process transport: + +```text +semantic VLLMError + -> normalize to OmniErrorDetails + -> stage/process transport + -> one boundary renderer + -> HTTP / SSE / WebSocket / async job / offline Python +``` + +The supported vLLM baseline must first contain +[vLLM PR #49665](https://github.com/vllm-project/vllm/pull/49665), including +commit `58f96593971b6287cdc8867024a6dfb9985c545c`. vLLM-Omni must import, not copy, +the upstream hierarchy. + +### 1. Use semantic exception ownership + +Generic failures use upstream classes directly: + +| Failure | Exception | Status | +| --- | --- | ---: | +| Invalid or incompatible request parameter | `VLLMValidationError` | 400 | +| Missing model, LoRA, voice, job, or artifact | `VLLMNotFoundError` | 404 | +| Permanently inaccessible or undecodable request media | `VLLMUnprocessableEntityError` | 422 | +| Recoverable per-request engine failure | `EngineGenerateError` | 500 | +| Unrecoverable engine failure | `EngineDeadError` | 500 | +| Unknown server failure | `VLLMServerError` or the `Exception` safety net | 500 | + +Omni adds only types with concrete Omni behavior: + +| Omni type | Parent | Status | Use | +| --- | --- | ---: | --- | +| `GuardrailViolationError` | `VLLMClientError` | 400 | Guardrail rejects request content | +| `OmniConflictError` | `VLLMClientError` | 409 | Existing async job is in an incompatible state | +| `OmniNotImplementedError` | `VLLMServerError` | 501 | Invoked server operation is not implemented | +| `OmniServiceUnavailableError` | `VLLMServerError` | 503 | Required handler, pipeline, or stage is unavailable | +| `OmniStageTimeoutError` | `VLLMServerError` | 504 | Stage or generation deadline expires | +| `OmniEngineGenerateError` | `EngineGenerateError` | 500 | Nonfatal stage failure with `stage_id` | +| `OmniEngineDeadError` | `EngineDeadError` | 500 | Fatal stage failure with `stage_id` | + +Each new subtype lands with its first production raise site. Existing 403, 413, +and 429 transport behavior remains compatible; dedicated types are added only +when a concrete use case and retry policy are defined. + +`OmniClientError` remains temporarily as a compatibility and remote-error +wrapper, but is reparented from `ValueError` to `VLLMClientError`. A symmetric +`OmniServerError(VLLMServerError)` represents server failures reconstructed +from the wire. + +### 2. Preserve semantics across stages + +Introduce one internal serializable record: + +```python +@dataclass(frozen=True, slots=True) +class OmniErrorDetails: + message: str + category: Literal["client", "server"] + status_code: int + error_type: str + parameter: str | None = None + request_id: str | None = None + stage_id: int | None = None + fatal: bool = False +``` + +Required invariants: + +- client errors have a 4xx status and server errors have a 5xx status; +- unknown exceptions normalize to server error 500; +- `fatal=True` is valid only for server errors; +- public messages and parameter names are sanitized; +- tracebacks, exception objects, media, tensors, credentials, and rejected + values are never serialized; and +- consumers use `category`, not the numeric status range, to reconstruct + ownership. + +`ErrorMessage`, `OmniRequestOutput`, `DiffusionOutput`, and stage IPC carry this +record or its complete field set. Existing flat message/status/type fields +remain readable for one compatibility window. + +### 3. Centralize rendering + +Register one Omni-aware `VLLMError` handler that: + +- preserves upstream status, type, and parameter mapping; +- dispatches engine errors through Omni's stage-aware termination behavior; +- renders Omni-specific 409, 501, 503, and 504 errors centrally; and +- runs inside the instrumented middleware boundary so HTTP metrics observe the + actual response status. + +Temporary handlers for raw `ValueError`, `TypeError`, `OverflowError`, and +`NotImplementedError` remain during migration. The permanent `Exception` +safety net always maps unknown failures to 500. + +Transport rules: + +- non-streaming HTTP status must equal `error.code`; +- SSE and WebSocket streams that already sent success headers emit one terminal + event with the canonical nested `error` object; +- async jobs store the canonical error and do not guess a new status when + retrieved; and +- offline APIs raise the semantic exception or a remote + `OmniClientError`/`OmniServerError` carrying `.details`. + +Routes must not return a Pydantic `ErrorResponse` without an explicit HTTP +status. + +### 4. Migrate incrementally + +1. **Baseline:** add table-driven mapping, middleware metrics, transport, and + engine-lifecycle regression tests for current behavior. +2. **Bridge:** adopt the required vLLM baseline, reparent compatibility types, + add `OmniErrorDetails`, and register the semantic handler. +3. **Request boundaries:** migrate project-owned validation one raise site at a + time to `VLLMValidationError`, `VLLMNotFoundError`, or + `VLLMUnprocessableEntityError`. Catch `VLLMClientError`, not `ValueError`. +4. **Stage and transports:** preserve stage context, standardize HTTP/SSE/ + WebSocket/job rendering, then remove raw fallbacks only when their audited + inventories reach zero or an explicit allowlist. + +This is not a mechanical replacement of every `ValueError`. Each site must be +reviewed to distinguish user-correctable input from internal failure. + +## Compatibility and correctness + +Correctly typed errors retain their current status and OpenAI-compatible +envelope. Intentional corrections may change: + +- an internal raw `ValueError` from 400 to 500; +- invalid or inaccessible media from 400/500 to 422; +- missing resources to 404; +- conflicting async-job operations to 409; and +- unavailable pipelines or stages to 503. + +Every status change requires an endpoint regression test and release note. + +Reparenting `OmniClientError` means external Python code that catches only +`ValueError` will no longer catch it. Retain its import and constructor for at +least one release, document the inheritance change, and recommend catching +`VLLMClientError` or a specific subtype. + +Minimum test coverage: + +- every upstream and Omni exception-to-status/type mapping; +- unknown `ValueError` becoming 500 after its fallback is removed; +- wire round trips preserving category, parameter, stage, and fatality; +- HTTP status matching `error.code`; +- SSE, WebSocket, and async-job terminal errors; +- `http_requests_total` distinguishing 4xx and 5xx; +- nonfatal stage failure affecting one request; +- fatal stage failure retaining existing termination behavior; and +- public errors excluding tracebacks, paths, secrets, and multimodal payloads. + +## Non-goals + +- Copying or forking upstream generic exception classes. +- Converting every model/kernel exception in one change. +- Making control-flow cancellation a public server error. +- Redesigning endpoint ownership, retry policy, or fatal OOM lifecycle. +- Adding speculative exception subclasses without a production raise site. + +## Open questions + +1. Should the generic 409, 501, 503, and 504 types eventually move upstream + after Omni validates their use cases? +2. Should `request_id` and `stage_id` become shared upstream response fields? +3. How long must mixed-version stage processes support the legacy wire fields? +4. For failed async jobs, should metadata retrieval return the stored error + status or HTTP 200 with a canonical embedded error? + +## Community help wanted + +We welcome community help with: + +- validating the proposed status and exception mapping; +- identifying representative raw-exception sites for the first migrations; +- reviewing mixed-version and multi-node wire compatibility; +- adding mapping, middleware-metrics, stage round-trip, and transport tests; and +- coordinating the required vLLM baseline update. + +If you would like to help, please comment with the endpoint, stage, or test area +you want to own. After the feedback period, implementation can be split into +small, independently reviewable PRs following the migration phases above. + +## Feedback period and CC + +Feedback is requested for two weeks after the RFC issue is published. + +CC: maintainers for entrypoints, engine/orchestrator, diffusion, metrics, and +the upstream vLLM error hierarchy. + +## References + +- [vLLM PR #44303: Fix 4xx errors recorded as 5xx in `http_requests_total`](https://github.com/vllm-project/vllm/pull/44303) +- [vLLM RFC #48227: Standardize vLLM Entrypoint Error Handling](https://github.com/vllm-project/vllm/issues/48227) +- [vLLM PR #49665: Standardize request error handling with `VLLMError`](https://github.com/vllm-project/vllm/pull/49665) +- [vLLM-Omni PR #3969: Correct HTTP status for voice errors](https://github.com/vllm-project/vllm-omni/pull/3969) +- [vLLM-Omni PR #4297: Guardrail client-error propagation](https://github.com/vllm-project/vllm-omni/pull/4297) +- [vLLM-Omni RFC #1346: Exit on OOM](https://github.com/vllm-project/vllm-omni/issues/1346) +- [vLLM-Omni RFC #5198: Broad exception cleanup](https://github.com/vllm-project/vllm-omni/issues/5198) +- [vLLM-Omni RFC #5227: Entrypoint and API-server refactor](https://github.com/vllm-project/vllm-omni/issues/5227)