651a88c09a · tduncan/Nvidia-Optimization-Stack-VLLM

benchmarks/benchmark_hidden_state_extraction.py

+1 -2

@@ -136,8 +136,7 @@ async def _client_loop(

136 client_id: int,

137 ):

138 """A single async client: pulls prompts, submits to engine, consumes

139 - hidden states as soon as each request finishes.

140 - """

139 + hidden states as soon as each request finishes."""

140 loop = asyncio.get_event_loop()

141 while True:

142 item = await prompt_queue.get()

benchmarks/fused_kernels/merge_attn_states_benchmarks.py

+1 -2

@@ -84,8 +84,7 @@ def make_inputs(

84

85 def build_configs(head_configs, num_tokens_list, input_dtypes, tp_sizes):

86 """Build (num_tokens, num_heads, head_size, dtype_str) config tuples,

87 - applying TP division to num_heads and skipping invalid combos.

88 - """

87 + applying TP division to num_heads and skipping invalid combos."""

88 configs = []

89 for (_, nh, hs), nt, dtype, tp in itertools.product(

90 head_configs, num_tokens_list, input_dtypes, tp_sizes

benchmarks/kernels/benchmark_moe_defaults.py

+1 -2

@@ -32,8 +32,7 @@ FP8_DTYPE = current_platform.fp8_dtype()

32

33 def old_default_config(M, E, N, K, topk, dtype=None, block_shape=None):

34 """The original defaults before https://github.com/vllm-project/vllm/pull/34846,

35 - for comparison.

36 - """

35 + for comparison."""

36 if dtype == "fp8_w8a8" and block_shape is not None:

37 return {

38 "BLOCK_SIZE_M": 64,

benchmarks/kernels/benchmark_vocab_parallel_embedding.py

+1 -2

@@ -97,8 +97,7 @@ def graph_replay(fn):

97

98 def calculate_diff(vocab_size, hidden_size, tp_size, num_tokens):

99 """The fused kernel must be bit-exact with the path it replaces, on every

100 - rank of the shard.

101 - """

100 + rank of the shard."""

101 dtype = torch.bfloat16

102 for tp_rank in range(tp_size):

103 input_ids, weight, shard_indices = make_inputs(

docs/mkdocs/gen_files/generate_examples.py

+2 -4

@@ -75,8 +75,7 @@ class Example:

75

76 If path is a file, it returns the path itself. If path is a directory, it

77 searches for Markdown files (*.md) in the directory and returns the first one

78 - found. If no Markdown files are found, it returns None.

79 - """

78 + found. If no Markdown files are found, it returns None."""

79 # Single file example

80 if self.path.is_file():

81 return self.path

@@ -91,8 +90,7 @@ class Example:

90 """Determine other files in the directory excluding the main file.

91

92 If path is a file, it returns an empty list. Otherwise, it returns every file

94 - in the directory except the main file in a list.

95 - """

93 + in the directory except the main file in a list."""

94 # Single file example

95 if self.path.is_file():

96 return []

examples/features/logits_processor/custom_req.py

+2 -4

@@ -55,8 +55,7 @@ logger = init_logger(__name__)

55

56 class DummyPerReqLogitsProcessor:

57 """The request-level logits processor masks out all logits except the

58 - token id identified by `target_token`

59 - """

58 + token id identified by `target_token`"""

59

60 def __init__(self, target_token: int) -> None:

61 """Specify `target_token`."""

@@ -75,8 +74,7 @@ class DummyPerReqLogitsProcessor:

74

75 class WrappedPerReqLogitsProcessor(AdapterLogitsProcessor):

76 """Example of wrapping a fake request-level logit processor to create a

78 - batch-level logits processor

79 - """

77 + batch-level logits processor"""

78

79 @classmethod

80 def validate_params(cls, params: SamplingParams):

examples/features/logits_processor/custom_req_init.py

+2 -4

@@ -56,8 +56,7 @@ logger = init_logger(__name__)

56

57 class DummyPerReqLogitsProcessor:

58 """The request-level logits processor masks out all logits except the

59 - token id identified by `target_token`

60 - """

59 + token id identified by `target_token`"""

60

61 def __init__(self, target_token: int) -> None:

62 """Specify `target_token`."""

@@ -76,8 +75,7 @@ class DummyPerReqLogitsProcessor:

75

76 class WrappedPerReqLogitsProcessor(AdapterLogitsProcessor):

77 """Example of overriding the wrapper class `__init__()` in order to utilize

79 - info about the device type

80 - """

78 + info about the device type"""

79

80 @classmethod

81 def validate_params(cls, params: SamplingParams):

examples/rl/rlhf_sharded_rdt_small_ep.py

+1 -2

@@ -101,8 +101,7 @@ PROMPTS = ["The capital of France is", "The future of AI is"]

101 class FSDPTrainWorker:

102 """One FSDP2 rank per GPU. Identical to the 8-GPU example's worker: the

103 sharded-RDT engine owns the NIXL serve surface, so this stays a plain Ray

104 - actor with no producer mixin.

105 - """

104 + actor with no producer mixin."""

105

106 def __init__(self, model_name, rank, world_size, master_addr, master_port):

107 self.rank = rank

pyproject.toml

+3 -1

@@ -90,7 +90,9 @@ ignore = [

90 # don't enforce that everything has a docstring

91 "D100", "D101", "D102", "D103", "D104", "D105", "D106", "D107",

92 # unwanted pydocstyle rules

93 - "D205", "D301", "D400", "D401", "D404", "D415",

93 + "D205", "D209", "D301", "D400", "D401", "D404", "D415",

94 + # incompatible pydocstyle pairs; pin the half `ruff` already picks

95 + "D203", "D213",

96 # star imports

97 "F405", "F403",

98 # lambda expression assignment

setup.py

+1 -2

@@ -724,8 +724,7 @@ class precompiled_wheel_utils:

724 wheels: list[dict], repo_url: str, arch: str

725 ) -> None:

726 """Warn if installed torch differs from the custom ROCm build on

727 - wheels.vllm.ai and suggest the correct install command.

728 - """

727 + wheels.vllm.ai and suggest the correct install command."""

728 try:

729 installed = torch.__version__

730 except Exception:

tests/basic_correctness/test_mem.py

+1 -2

@@ -112,8 +112,7 @@ def test_basic_cumem():

112 @create_new_process_for_each_test("fork" if current_platform.is_cuda() else "spawn")

113 def test_discard_tags():

114 """Test that discard(tags) selectively frees GPU memory for specific

115 - tags while keeping other tags mapped and usable.

116 - """

115 + tags while keeping other tags mapped and usable."""

116 allocator = get_mem_allocator_instance()

117

118 with allocator.use_memory_pool("weights"):

tests/benchmarks/test_bfcl_dataset.py

+6 -12

@@ -13,8 +13,7 @@ from vllm.benchmarks.datasets import BFCLDataset, get_samples

13

14 def _patch_hf_api(side_effect):

15 """Return a patch context that swaps `hf_api()` to a stub whose

16 - `.hf_hub_download` attribute uses `side_effect`.

17 - """

16 + `.hf_hub_download` attribute uses `side_effect`."""

17 fake_api = MagicMock()

18 fake_api.hf_hub_download.side_effect = side_effect

19 return patch("vllm.benchmarks.datasets.datasets.hf_api", return_value=fake_api)

@@ -116,8 +115,7 @@ def test_bfcl_dataset_translates_schema_and_attaches_tools(

115 ) -> None:

116 """BFCLDataset should translate schemas to OpenAI tool format, set

117 `messages` directly on SampleRequest, and attach tools/tool_choice via

119 - request_overrides.

120 - """

118 + request_overrides."""

119 paths = _write_fake_files(tmp_path)

120

121 def fake_download(_repo, filename, **_kwargs):

@@ -168,8 +166,7 @@ def test_bfcl_dataset_missing_category_raises_clear_error(

166 hf_tokenizer: PreTrainedTokenizerBase,

167 ) -> None:

168 """A typo'd category should produce an actionable ValueError, not an

171 - opaque huggingface_hub exception.

172 - """

169 + opaque huggingface_hub exception."""

170 from huggingface_hub.errors import EntryNotFoundError

171

172 args = _args_for_bfcl(categories=["simpl"]) # typo

@@ -187,8 +184,7 @@ def test_bfcl_dataset_missing_category_raises_clear_error(

184 @pytest.mark.benchmark

185 def test_chat_backend_uses_messages_field_when_set() -> None:

186 """When RequestFuncInput.chat_messages is set, the chat backend must use

190 - it verbatim and skip default content construction from `prompt`.

191 - """

187 + it verbatim and skip default content construction from `prompt`."""

188 import asyncio

189

190 from vllm.benchmarks.lib.endpoint_request_func import (

@@ -248,8 +244,7 @@ def test_chat_backend_uses_messages_field_when_set() -> None:

244 def test_bfcl_prompt_len_includes_tools(tmp_path: Path) -> None:

245 """prompt_len must reflect tokens from both messages *and* tool schemas,

246 so percentile buckets and input-distribution summaries aren't biased

251 - low for tool-heavy traffic.

252 - """

247 + low for tool-heavy traffic."""

248 paths = _write_fake_files(tmp_path)

249

250 def fake_download(_repo, filename, **_kwargs):

@@ -296,8 +291,7 @@ def test_bfcl_prompt_len_falls_back_when_tokenizer_rejects_tools(

291 tmp_path: Path,

292 ) -> None:

293 """Older tokenizers don't accept tools=; fallback must still produce a

299 - non-zero prompt_len without crashing.

300 - """

294 + non-zero prompt_len without crashing."""

295 paths = _write_fake_files(tmp_path)

296

297 def fake_download(_repo, filename, **_kwargs):

tests/benchmarks/test_custom_dataset_chat_template_kwargs.py

+1 -2

@@ -12,8 +12,7 @@ from vllm.benchmarks.datasets import get_samples

12 class _RecordingTokenizer:

13 """Minimal tokenizer stub that records the kwargs forwarded to

14 apply_chat_template, so we can assert chat_template_kwargs propagation

15 - without loading a real model/template.

16 - """

15 + without loading a real model/template."""

16

17 def __init__(self) -> None:

18 self.captured_kwargs: dict | None = None

tests/benchmarks/test_sampling_params.py

+3 -6

@@ -10,8 +10,7 @@ from vllm.tokenizers import TokenizerLike

10

11 class _FakeTokenizer(TokenizerLike):

12 """Minimal tokenizer implementing the TokenizerLike protocol

13 - for testing get_sampling_params.

14 - """

13 + for testing get_sampling_params."""

14

15 def __init__(self, vocab_size: int = 1000, num_special_tokens: int = 0) -> None:

16 self._vocab_size = vocab_size

@@ -172,8 +171,7 @@ class TestGetSamplingParams:

171

172 def test_output_low_clamped_to_one(self):

173 """Even with a high ratio that would push output_low to 0,

175 - the function clamps it to 1.

176 - """

174 + the function clamps it to 1."""

175 rng = np.random.default_rng(0)

176 # output_len=1, ratio=0.99 → floor(1*0.01)=0, should clamp to 1

177 _, output_lens, _ = get_sampling_params(

@@ -235,8 +233,7 @@ class TestGetSamplingParams:

233

234 def test_input_len_zero_with_special_tokens(self):

235 """input_len < num_special_tokens → real_input_len = 0, which is fine

238 - (range [0, 0]).

239 - """

236 + (range [0, 0])."""

237 rng = np.random.default_rng(0)

238 input_lens, _, _ = get_sampling_params(

239 rng, 5, 0.0, 5, 50, self._tok(num_special=10)

tests/benchmarks/test_skip_tokenizer_init.py

+1 -2

@@ -52,8 +52,7 @@ def _write_dataset(path: Path) -> None:

52

53 def _args(dataset_path: str) -> argparse.Namespace:

54 """Reproduce the argparse.Namespace that serve.py builds from the

55 - failing command, including skip_tokenizer_init=True.

56 - """

55 + failing command, including skip_tokenizer_init=True."""

56 return argparse.Namespace(

57 # dataset

58 dataset_name="custom",

tests/compile/passes/test_fusion.py

+1 -2

@@ -732,8 +732,7 @@ def test_aiter_fusion_rmsnorm_gated_quant_no_gdn_layers(

732 monkeypatch: pytest.MonkeyPatch,

733 ):

734 """Verify that without GDN layers in static_forward_context,

735 - the gated pattern is not registered and no matches occur.

736 - """

735 + the gated pattern is not registered and no matches occur."""

736 group_shape = GroupShape(1, 128)

737 vllm_config = VllmConfig(

738 model_config=ModelConfig(dtype=dtype),

tests/compile/passes/test_split_coalescing.py

+1 -2

@@ -15,8 +15,7 @@ DEVICE_TYPE = current_platform.device_type

15

16 class SplitCoalescingModel(torch.nn.Module):

17 """Model with 3 separate split_with_sizes calls on the same input,

18 - simulating the B200+FP8 graph where CSE fails to merge them.

19 - """

18 + simulating the B200+FP8 graph where CSE fails to merge them."""

19

20 def __init__(self, q_size: int, kv_size: int) -> None:

21 super().__init__()

tests/compile/test_codegen.py

+10 -20

@@ -77,8 +77,7 @@ def test_non_primitive_kwargs_lifted_to_consts(

77 into the namespace, so ``device`` is unbound. The fix collects such

78 objects into ``__vllm_consts__`` and references them by index. The

79 unqualified ``device(type=...)`` form must never appear in the

80 - generated source.

81 - """

80 + generated source."""

81 split_gm = _trace_and_split(model_fn, (x,), split_ops)

82 code, submod_names, consts = generate_execution_code(split_gm)

83

@@ -99,8 +98,7 @@ def test_dtype_singleton_deduped(x: torch.Tensor) -> None:

98 """``torch.float16`` is a process-wide singleton, so two ops referring

99 to it in the traced graph share a single consts slot via ``id()``-based

100 dedup. Distinct expressions (``x.to(...)`` vs ``(x*2).to(...)``) ensure

102 - the tracer can't CSE the two ops into a single node.

103 - """

101 + the tracer can't CSE the two ops into a single node."""

102

103 def model_fn(x: torch.Tensor) -> torch.Tensor:

104 return x.to(dtype=torch.float16) + (x * 2).to(dtype=torch.float16)

@@ -149,8 +147,7 @@ def test_distinct_dtypes_get_distinct_slots(x: torch.Tensor) -> None:

147 def test_consts_ordering_deterministic(x: torch.Tensor) -> None:

148 """Two independent traces of the same model must produce equal consts

149 lists *in the same order*. Cache artifacts identify const slots by

152 - index, so a non-deterministic order would invalidate cached code.

153 - """

150 + index, so a non-deterministic order would invalidate cached code."""

151

152 def model_fn(x: torch.Tensor) -> torch.Tensor:

153 # Multiple distinct non-primitives encountered in a fixed graph order.

@@ -188,8 +185,7 @@ def test_consts_shared_across_split_submods(x: torch.Tensor) -> None:

185 The function below splits into three inlined submods, two of which

186 independently reference ``torch.float16``. The shared ``const_index``

187 threaded through recursive ``generate_execution_code_with_name`` calls

191 - must collapse the dtype to a single slot used from both submods.

192 - """

188 + must collapse the dtype to a single slot used from both submods."""

189

190 def model_fn(x: torch.Tensor) -> torch.Tensor:

191 a = x.to(dtype=torch.float16) # submod_0: _to_copy(fp16)

@@ -234,8 +230,7 @@ def test_non_graphmodule_submod_uses_indexed_callable(x: torch.Tensor) -> None:

230 """When a child of split_gm is *not* a ``torch.fx.GraphModule`` — as

231 happens in production once ``PiecewiseBackend`` replaces submods —

232 codegen emits ``__vllm_submods__[idx](...)`` instead of inlining, and

237 - the runtime callable is bound from ``submod_callables``.

238 - """

233 + the runtime callable is bound from ``submod_callables``."""

234

235 def model_fn(x: torch.Tensor) -> torch.Tensor:

236 return x.relu().sigmoid()

@@ -285,8 +280,7 @@ def test_non_graphmodule_submod_uses_indexed_callable(x: torch.Tensor) -> None:

280 def test_getitem_in_stitching_graph(x: torch.Tensor) -> None:

281 """``operator.getitem`` on submod tuple returns is the ``call_function``

282 special case at codegen.py — emitted as ``name = source[index]``

288 - rather than a function call.

289 - """

283 + rather than a function call."""

284

285 def model_fn(x: torch.Tensor) -> torch.Tensor:

286 return x.relu().sigmoid()

@@ -304,8 +298,7 @@ def test_getitem_in_stitching_graph(x: torch.Tensor) -> None:

298 def test_del_emitted_for_intermediate_values(x: torch.Tensor) -> None:

299 """The codegen schedules ``del`` after a value's last use to free

300 memory early. Multi-submod splits naturally have intermediates whose

307 - last use is not the output node.

308 - """

301 + last use is not the output node."""

302

303 def model_fn(x: torch.Tensor) -> torch.Tensor:

304 return x.relu().sigmoid().tanh()

@@ -325,8 +318,7 @@ def test_with_submod_false_rejects_call_module() -> None:

318 """``generate_execution_code_with_name(with_submod=False)`` is the

319 recursive entry for inlining a GraphModule into its parent. It must

320 refuse a graph that itself contains ``call_module`` nodes — the parent

328 - is responsible for handling those.

329 - """

321 + is responsible for handling those."""

322 g = fx.Graph()

323 x_node = g.placeholder("x")

324 root = torch.nn.Module()

@@ -343,8 +335,7 @@ def test_node_ref_recurses_through_containers() -> None:

335 """``_node_ref`` is the recursive walker that lifts non-primitives

336 nested inside list/tuple/dict args. Real aten ops rarely produce such

337 structures, but the path is needed for DTensor placement lists and

346 - other future cases — unit-test the walker directly.

347 - """

338 + other future cases — unit-test the walker directly."""

339 consts: list = []

340 const_index: dict[int, int] = {}

341 cpu = torch.device("cpu")

@@ -368,8 +359,7 @@ def test_node_ref_recurses_through_containers() -> None:

359 def test_legacy_code_without_consts() -> None:

360 """``compile_execution_fn(consts=None)`` must still load code that has

361 no ``__vllm_consts__`` reference, so older serialized cache artifacts

371 - keep working.

372 - """

362 + keep working."""

363 # Pre-consts codegen: no __vllm_consts__ reference, only torch/operator.

364 legacy_code = (

365 "import torch\n"

tests/compile/test_compile_ranges.py

+2 -4

@@ -129,8 +129,7 @@ def test_compile_config_get_compile_ranges():

129

130 class PostGradStaticShapeChecker(InductorPass):

131 """Asserts that compile_sizes entries produce graphs with fully concrete

132 - (non-symbolic) shapes, and compile_ranges entries have symbolic shapes.

133 - """

132 + (non-symbolic) shapes, and compile_ranges entries have symbolic shapes."""

133

134 def __init__(self):

135 self.num_static_calls = 0

@@ -171,8 +170,7 @@ class PostGradStaticShapeChecker(InductorPass):

170

171 def test_compile_sizes_produce_static_shapes(disable_vllm_compile_cache):

172 """Verify that compile_sizes entries are compiled with fully concrete

174 - shapes (no SymInts), while compile_ranges entries retain dynamic shapes.

175 - """

173 + shapes (no SymInts), while compile_ranges entries retain dynamic shapes."""

174 checker = PostGradStaticShapeChecker()

175 torch.set_default_device("cuda")

176 vllm_config = VllmConfig(

tests/compile/test_config.py

+3 -6

@@ -1229,8 +1229,7 @@ def test_compile_sizes_padding_validation():

1229

1230 def test_inductor_asserts_default_disabled(monkeypatch):

1231 """Test that inductor runtime asserts are disabled by default

1232 - (INFO logging level) on torch < 2.12.

1233 - """

1232 + (INFO logging level) on torch < 2.12."""

1233 monkeypatch.setenv("VLLM_LOGGING_LEVEL", "INFO")

1234

1235 import importlib

@@ -1248,8 +1247,7 @@ def test_inductor_asserts_default_disabled(monkeypatch):

1247

1248 def test_inductor_asserts_enabled_in_debug(monkeypatch):

1249 """Test that VLLM_LOGGING_LEVEL=DEBUG enables inductor runtime asserts

1251 - on torch < 2.12.

1252 - """

1250 + on torch < 2.12."""

1251 monkeypatch.setenv("VLLM_LOGGING_LEVEL", "DEBUG")

1252

1253 import importlib

@@ -1285,8 +1283,7 @@ def test_get_inductor_factors_includes_configs():

1283

1284 def test_inductor_asserts_user_override(monkeypatch):

1285 """Test that explicit inductor_compile_config overrides the

1288 - debug-logging default.

1289 - """

1286 + debug-logging default."""

1287 monkeypatch.setenv("VLLM_LOGGING_LEVEL", "INFO")

1288

1289 import importlib

tests/compile/test_inductor_fallback_allow_list_patch.py

+1 -2

@@ -236,8 +236,7 @@ class TestInductorFallbackFastPath:

236

237 def test_inner_set_membership_still_works_for_standard_ops(self):

238 """The patch must not break Inductor's existing fallback decisions

239 - for non-vllm ops such as ``torchvision::roi_align``.

240 - """

239 + for non-vllm ops such as ``torchvision::roi_align``."""

240 import torch._inductor.lowering as _lowering

241

242 allow_list = _lowering.FALLBACK_ALLOW_LIST

tests/config/test_model_arch_config.py

+2 -4

@@ -372,8 +372,7 @@ class _HeterogeneousConfig(PretrainedConfig):

372 """A heterogeneous config with no convertor of its own.

373

374 Mirrors the parts vLLM uses: per-layer configs are shallow copies with the

375 - varying attributes applied and heterogeneity stripped, so they do not recurse.

376 - """

375 + varying attributes applied and heterogeneity stripped, so they do not recurse."""

376

377 is_heterogeneous = True

378

@@ -472,8 +471,7 @@ def test_deepseek_v4_convertor_splits_vision_architecture():

471 """The vision checkpoint shares model_type/architectures with the

472 text-only DeepSeek-V4; the convertor routes it to the VL wrapper class

473 by rewriting hf_config.architectures (model-class resolution reads the

475 - raw attribute).

476 - """

474 + raw attribute)."""

475 from vllm.transformers_utils.configs.deepseek_v4 import DeepseekV4Config

476 from vllm.transformers_utils.model_arch_config_convertor import (

477 DeepseekV4ModelArchConfigConvertor,

tests/config/test_multimodal_config.py

+5 -10

@@ -42,8 +42,7 @@ def test_mm_encoder_attn_backend_hash_updates():

42

43 def test_language_model_only_does_not_affect_mm_hash():

44 """language_model_only does not affect the ViT computation graph,

45 - so it should not change the multimodal config hash.

46 - """

45 + so it should not change the multimodal config hash."""

46 base_hash = MultiModalConfig().compute_hash()

47 lm_only_hash = MultiModalConfig(language_model_only=True).compute_hash()

48 assert base_hash == lm_only_hash

@@ -51,8 +50,7 @@ def test_language_model_only_does_not_affect_mm_hash():

50

51 def test_language_model_only_affects_model_hash():

52 """language_model_only affects the LM computation graph,

54 - so it should change the model config hash.

55 - """

53 + so it should change the model config hash."""

54 model = "llava-hf/llava-1.5-7b-hf"

55 base_hash = ModelConfig(model).compute_hash()

56 lm_only_hash = ModelConfig(model, language_model_only=True).compute_hash()

@@ -388,8 +386,7 @@ def _resolve_mm_video_decode_device(

386 torchcodec_available: bool = True,

387 ) -> dict:

388 """Run processor-device then video-decode-device resolution and report

391 - the resulting video media IO kwargs.

392 - """

389 + the resulting video media IO kwargs."""

390 mm_config = MultiModalConfig(

391 mm_processor_kwargs={} if device is None else {"device": device},

392 mm_tensor_ipc=mm_tensor_ipc, # type: ignore[arg-type]

@@ -419,8 +416,7 @@ def _resolve_mm_video_decode_device(

416

417 def test_auto_video_decode_uses_nvdec_on_encoder_instance():

418 """The processor runs on the accelerator there, so decoded frames should

422 - stay on-device too.

423 - """

419 + stay on-device too."""

420 assert _resolve_mm_video_decode_device(ec_role="ec_producer") == {

421 "backend": "torchcodec",

422 "device": "cuda",

@@ -432,8 +428,7 @@ def test_auto_video_decode_stays_on_cpu_off_encoder_instance(

428 ec_role: ECRole | None,

429 ):

430 """The processor stays on CPU off encode-only instances, so video

435 - decoding should too.

436 - """

431 + decoding should too."""

432 assert _resolve_mm_video_decode_device(ec_role=ec_role) == {}

433

434

tests/config/test_speculative_draft_hf_overrides.py

+4 -8

@@ -34,8 +34,7 @@ def _make_hf_config(**kwargs) -> PretrainedConfig:

34 @pytest.mark.cpu_test

35 def test_dict_overrides_are_not_forwarded_to_draft():

36 """Dict overrides are target-specific key patches; the draft must get

37 - only the architecture-mapping override.

38 - """

37 + only the architecture-mapping override."""

38 composed = SpeculativeConfig.compose_draft_hf_overrides(

39 {"max_position_embeddings": 1234}

40 )

@@ -51,8 +50,7 @@ def test_none_overrides_fall_back_to_arch_mapping():

50 @pytest.mark.cpu_test

51 def test_callable_overrides_reach_the_draft_config():

52 """A callable override (config-to-config transform) composes with the

54 - architecture-mapping override and is applied to the draft config.

55 - """

53 + architecture-mapping override and is applied to the draft config."""

54

55 def shrink(hf_config: PretrainedConfig) -> PretrainedConfig:

56 hf_config.num_hidden_layers = 1

@@ -69,8 +67,7 @@ def test_callable_overrides_reach_the_draft_config():

67 @pytest.mark.cpu_test

68 def test_arch_mapping_applies_before_callable_override():

69 """The static arch-mapping override runs first, so the user callable

72 - observes (and may adjust) the post-mapping config.

73 - """

70 + observes (and may adjust) the post-mapping config."""

71 seen_architectures: list[str] = []

72

73 def record(hf_config: PretrainedConfig) -> PretrainedConfig:

@@ -131,8 +128,7 @@ def test_composed_override_is_picklable():

128 the composed override must be picklable. A nested local closure is not

129 (it raised ``Can't get local object`` on DFlashDraftModel); a

130 ``functools.partial`` over a module-referenceable static method is.

134 - Guard against regressing to a closure.

135 - """

131 + Guard against regressing to a closure."""

132 composed = SpeculativeConfig.compose_draft_hf_overrides(_module_level_shrink)

133

134 assert isinstance(composed, functools.partial)

tests/config/test_speculative_draft_max_position_embeddings.py

+1 -2

@@ -99,8 +99,7 @@ def test_independent_draft_model_keeps_its_own_limit(

99 vllm_caplog: pytest.LogCaptureFixture,

100 ):

101 """An independent AR draft may genuinely have a smaller context than the

102 - target; its max_position_embeddings must not be resized.

103 - """

102 + target; its max_position_embeddings must not be resized."""

103 target_model_config = ModelConfig(

104 AR_MODEL, hf_overrides={"max_position_embeddings": 8192}

105 )

tests/distributed/conftest.py

+1 -2

@@ -179,8 +179,7 @@ class MockSubscriber:

179 @pytest.fixture

180 def enable_ray_v2_backend():

181 """Set env vars for the Ray V2 executor backend and shut down Ray

182 - between tests.

183 - """

182 + between tests."""

183 import ray

184

185 saved = {

tests/distributed/test_eplb_quant_scale_consistency.py

+4 -8

@@ -124,8 +124,7 @@ def _build_processed_layer(

124 backend: str, raw: dict[str, torch.Tensor], device: torch.device

125 ) -> tuple[CompressedTensorsW4A4Nvfp4MoEMethod, _RoutedExpertsStub]:

126 """Create the method + layer, load ``raw`` and run the real

127 - process_weights_after_loading (which builds the flashinfer kernel).

128 - """

127 + process_weights_after_loading (which builds the flashinfer kernel)."""

128 moe_config = _make_moe_config(backend)

129 method = CompressedTensorsW4A4Nvfp4MoEMethod(moe_config, "layer.0", use_a16=False)

130 layer = _RoutedExpertsStub(moe_config).to(device)

@@ -158,8 +157,7 @@ def _assert_tensors_equal(

157

158 def _simulate_eplb_rearrangement(layer: torch.nn.Module, perm: torch.Tensor) -> None:

159 """Permute experts of every registered Parameter in place, the way

161 - EPLB's in-place rearrangement moves expert slices along dim 0.

162 - """

160 + EPLB's in-place rearrangement moves expert slices along dim 0."""

161 with torch.no_grad():

162 for _, param in layer.named_parameters():

163 param.copy_(param[perm])

@@ -167,8 +165,7 @@ def _simulate_eplb_rearrangement(layer: torch.nn.Module, perm: torch.Tensor) ->

165

166 def _ensure_world1_distributed() -> None:

167 """The enable_eplb config makes weight processing all-reduce activation

170 - scale amaxes over the EP group, so a (single-rank) EP group must exist.

171 - """

168 + scale amaxes over the EP group, so a (single-rank) EP group must exist."""

169 from vllm.distributed.parallel_state import (

170 ensure_model_parallel_initialized,

171 init_distributed_environment,

@@ -256,8 +253,7 @@ def test_nvfp4_eplb_rearrangement_matches_reload(backend: str) -> None:

253 def test_fp8_per_tensor_alphas_registered_and_aliased() -> None:

254 """CPU-only: the fp8 per-tensor oracle must register the fused

255 (w_scale * a_scale) products as layer Parameters aliased by the quant

259 - config, so EPLB rearrangement keeps kernels consistent.

260 - """

256 + config, so EPLB rearrangement keeps kernels consistent."""

257 from vllm.model_executor.layers.fused_moe.oracle.fp8 import (

258 Fp8MoeBackend,

259 make_fp8_moe_quant_config,

tests/distributed/test_mnnvl_alltoall.py

+2 -4

@@ -160,8 +160,7 @@ def _make_forward_context(rank, world_size, num_tokens_per_rank):

160 class _AttnMeta:

161 """Minimal placeholder so set_forward_context's

162 ``attn_metadata is not None`` guard (forward_context.py:334)

163 - is satisfied. The real DPMetadata is built from num_tokens_across_dp.

164 - """

163 + is satisfied. The real DPMetadata is built from num_tokens_across_dp."""

164

165 dp_metadata = None

166

@@ -460,8 +459,7 @@ def _one_sided_workspace_grow_worker(rank, world_size):

459 def test_one_sided_manager_workspace_grow(world_size):

460 """A later initialize() with a larger per-token payload must grow the

461 workspace and rebuild MoeAlltoAll; a later initialize() with a smaller

463 - payload must no-op.

464 - """

462 + payload must no-op."""

463 _spawn_workers(

464 _one_sided_workspace_grow_worker,

465 world_size,

tests/distributed/test_mq_connect_ip.py

+3 -6

@@ -21,8 +21,7 @@ from vllm.utils.network_utils import get_ip

21

22 def test_mq_bind_with_local_ip():

23 """MessageQueue with remote readers should successfully bind

24 - when connect_ip is the local node's IP.

25 - """

24 + when connect_ip is the local node's IP."""

25 # n_reader=2, n_local_reader=1 means 1 remote reader,

26 # which triggers the remote ZMQ socket bind.

27 mq = MessageQueue(

@@ -44,8 +43,7 @@ def test_mq_bind_with_local_ip():

43 def test_mq_bind_with_non_local_ip_fails():

44 """MessageQueue should fail to bind when connect_ip is a

45 non-local IP address (simulating the bug where master_addr

47 - from a different node was used).

48 - """

46 + from a different node was used)."""

47 # Use a non-local IP that we definitely can't bind to.

48 # 198.51.100.1 is from TEST-NET-2 (RFC 5737), never locally assigned.

49 non_local_ip = "198.51.100.1"

@@ -59,8 +57,7 @@ def test_mq_bind_with_non_local_ip_fails():

57

58 def test_mq_bind_defaults_to_local_ip():

59 """When connect_ip is None, MessageQueue should auto-detect

62 - the local IP and bind successfully.

63 - """

60 + the local IP and bind successfully."""

61 mq = MessageQueue(

62 n_reader=2,

63 n_local_reader=1,

tests/distributed/test_pynccl.py

+1 -2

@@ -495,8 +495,7 @@ def test_pynccl_suspend_resume_idempotent():

495

496 def test_pynccl_suspend_noop_without_symbol():

497 """RCCL / NCCL < 2.29.7 lack ncclCommSuspend; suspend() must no-op, not

498 - crash, and resume() must then no-op too since nothing was suspended.

499 - """

498 + crash, and resume() must then no-op too since nothing was suspended."""

499 from unittest.mock import Mock

500

501 comm = object.__new__(PyNcclCommunicator)

tests/distributed/test_ray_v2_executor_e2e.py

+1 -2

@@ -23,8 +23,7 @@ def _ray_init():

23

24 Without this, workers cannot unpickle actor classes defined in the

25 ``tests`` package, causing FunctionActorManager to fall back to

26 - TemporaryActor which drops async method signatures.

27 - """

26 + TemporaryActor which drops async method signatures."""

27 project_root = str(pathlib.Path(__file__).resolve().parents[2])

28 ray.init(

29 ignore_reinit_error=True,

tests/distributed/test_sharded_rdt_plan.py

+61 -122

@@ -53,8 +53,7 @@ META = torch.device("meta")

53

54 class _FakeLayer:

55 """Stand-in for a leaf module. The planner only stores it and keys on

56 - ``id()``; nothing in the plan path touches its attributes.

57 - """

56 + ``id()``; nothing in the plan path touches its attributes."""

57

58 def __init__(self, tag: str) -> None:

59 self.tag = tag

@@ -65,8 +64,7 @@ class _FakeLayer:

64

65 def _module(layer, copies):

66 """What the bake records for one leaf module: a scatter list whose entries

68 - all carry that module as their ``layer``.

69 - """

67 + all carry that module as their ``layer``."""

68 for c in copies:

69 c.layer = layer

70 return copies

@@ -83,8 +81,7 @@ def _copy(

81 dtype=torch.bfloat16,

82 ):

83 """A `_Scatter` as the bake would record it: source key, owning layer, and

86 - the meta destination region.

87 - """

84 + the meta destination region."""

85 stride = (

86 (1,)

87 if len(shape) == 1

@@ -143,8 +140,7 @@ def _planner(baked, *, name_meta, live=None, held_by=None, group_lens=None):

140

141 def _one_module_per_layer(n_layers, *, dtype="bfloat16", numel=4):

142 """`(baked, name_meta, names, group_lens)` for a pre / N-layer / post model

146 - where every gather group is one fully-baked leaf module.

147 - """

143 + where every gather group is one fully-baked leaf module."""

144 names = ["embed.weight"]

145 names += [f"model.layers.{i}.w" for i in range(n_layers)]

146 names += ["norm.weight"]

@@ -165,8 +161,7 @@ def _one_module_per_layer(n_layers, *, dtype="bfloat16", numel=4):

161 class TestFakeOpChains:

162 """Every allowlisted op must append itself to the chain and hand back a

163 child whose shape/dtype PyTorch itself computed. The chain is the wire

168 - format the producer replays, so its exact contents are load-bearing.

169 - """

164 + format the producer replays, so its exact contents are load-bearing."""

165

166 def _fake(self, shape=(4, 6), dtype=torch.bfloat16, sink=None):

167 return FakeRDTTensor(

@@ -227,8 +222,7 @@ class TestFakeOpChains:

222 @pytest.mark.parametrize("op,n", [("chunk", 2), ("unbind", 4)])

223 def test_multi_return_op_emits_one_child_per_output(self, op, n):

224 """chunk/unbind hand back a tuple; each child carries the base op plus a

230 - trailing __getitem__(i) so the producer can index the replayed result.

231 - """

225 + trailing __getitem__(i) so the producer can index the replayed result."""

226 parts = getattr(self._fake(), op)(*((n,) if op == "chunk" else ()), 0)

227 assert isinstance(parts, tuple)

228 assert len(parts) == n

@@ -244,8 +238,7 @@ class TestFakeOpChains:

238

239 class TestFakeUnsupportedOps:

240 """Anything that needs real data must fail loudly at bake time rather than

247 - silently transferring the wrong bytes.

248 - """

241 + silently transferring the wrong bytes."""

242

243 def _fake(self):

244 return FakeRDTTensor(

@@ -285,8 +278,7 @@ class TestFakeUnsupportedOps:

278

279 class TestBakeRecording:

280 """During the dry run the fake's ``copy_`` is the data sink: it records the

288 - source chain plus the meta destination's strided region and moves nothing.

289 - """

281 + source chain plus the meta destination's strided region and moves nothing."""

282

283 def _recorder_and_fake(self, shape=(4, 6)):

284 rec = BakeSink()

@@ -322,8 +314,7 @@ class TestBakeRecording:

314

315 def test_unattributed_copy_is_live_but_unrecorded(self):

316 """A copy_ with no loader stamp cannot be attributed to a param, so its

325 - module must fall back to the plain load — but the name still moved data.

326 - """

317 + module must fall back to the plain load — but the name still moved data."""

318 rec, fake = self._recorder_and_fake()

319 rec.current = None

320 torch.empty((4, 6), dtype=torch.bfloat16, device=META).copy_(fake)

@@ -363,8 +354,7 @@ class TestBakeRecording:

354 producer packs what the replay yields, so the record must carry the

355 POST-chain dtype: taking it from the source name's metadata instead sizes

356 the slice with the wrong itemsize and shifts every later slice in the

366 - chunk, carving the packed blob differently on the two sides.

367 - """

357 + chunk, carving the packed blob differently on the two sides."""

358 rec, fake = self._recorder_and_fake(shape=(4,))

359 layer = _FakeLayer("reinterpreted")

360 rec.current = (layer, "weight")

@@ -384,8 +374,7 @@ class TestBakeRecording:

374 the producer packs what the chain yields, and nothing downstream can

375 catch a mismatch — the consumer's buffer view is exactly prod(dest.shape)

376 elements, so it reshapes cleanly over bytes laid out at other offsets.

387 - So the bake refuses rather than recording a slice it cannot carve.

388 - """

377 + So the bake refuses rather than recording a slice it cannot carve."""

378 rec, fake = self._recorder_and_fake(shape=(1, 6))

379 rec.current = (_FakeLayer("broadcast"), "weight")

380 param = torch.empty((4, 6), dtype=torch.bfloat16, device=META)

@@ -401,8 +390,7 @@ class TestBakeRecording:

390 class TestConsumerIdentity:

391 """Every worker in the fleet needs a DISTINCT id in 0..C-1: it selects the

392 worker's producer block and keys the producer's per-consumer serve ring, so

404 - a collision silently serves two workers out of one ring.

405 - """

393 + a collision silently serves two workers out of one ring."""

394

395 def _engine(self, *, dp_index, rank, world_size):

396 eng = object.__new__(ShardedRDTWeightTransferEngine)

@@ -439,8 +427,7 @@ class TestConsumerIdentity:

427 def test_independent_engines_offset_into_distinct_ranges(self):

428 """A fleet of separate engines restarts _global_worker_index at 0 in

429 each, so without the replica offset every engine would claim 0..w-1.

442 - The driver sets replica_rank on the payload; these are the ids it buys.

443 - """

430 + The driver sets replica_rank on the payload; these are the ids it buys."""

431 ids = [

432 self._ids(num_consumers=8, workers=range(4), replica_rank=r, num_replicas=2)

433 for r in (0, 1)

@@ -466,8 +453,7 @@ def _router(

453 group_lens=None,

454 ):

455 """A router over ``n`` single-name groups unless told otherwise, so a test

469 - can state expectations in group indices.

470 - """

456 + can state expectations in group indices."""

457 if names is None:

458 n = len(name_owner_class or []) or (len(group_lens or []) or 1)

459 names = [f"g{i}" for i in range(n)]

@@ -496,8 +482,7 @@ class TestRdtRouter:

482 def test_every_producer_nic_carries_traffic(self):

483 """16 producers / 8 consumers: the block rule spreads each consumer's

484 pulls over its block, alternating by group, so no producer sits idle.

499 - Plain `consumer_id % len(owners)` would idle half of them.

500 - """

485 + Plain `consumer_id % len(owners)` would idle half of them."""

486 r = _router(16, 8, group_lens=[1] * 95)

487 names = [f"g{i}" for i in range(95)]

488 for c in range(8):

@@ -515,8 +500,7 @@ class TestRdtRouter:

500

501 def test_pipeline_stages_route_to_the_owning_stage(self):

502 """2 stages x 8 ranks: groups 0-2 on stage 0, 3-5 on stage 1. Two owner

518 - sets, one per stage; each consumer reaches both.

519 - """

503 + sets, one per stage; each consumer reaches both."""

504 r = _router(

505 16,

506 8,

@@ -532,8 +516,7 @@ class TestRdtRouter:

516

517 def test_expert_names_route_to_the_holding_ranks(self):

518 """A name held by only some ranks routes to those; a replicated name

535 - keeps the full set.

536 - """

519 + keeps the full set."""

520 r = RdtRouter(

521 8,

522 4,

@@ -552,8 +535,7 @@ class TestRdtRouter:

535

536 def test_consumers_spread_over_an_owner_set(self):

537 """Several ranks hold a name (its TP peers); the block rule must spread

555 - consumers across them, not funnel through one.

556 - """

538 + consumers across them, not funnel through one."""

539 r = _router(

540 4, 4, owner_sets=[[0, 1]], name_owner_class=[0] * 6, group_lens=[1] * 6

541 )

@@ -562,8 +544,7 @@ class TestRdtRouter:

544

545 def test_a_group_frees_at_every_owner_of_any_of_its_names(self):

546 """The free barrier is per group, so its fan-out is the UNION over the

565 - group's names: a rank holding none of them is not signalled.

566 - """

547 + group's names: a rank holding none of them is not signalled."""

548 r = RdtRouter(4, 2, [[0, 1], [2, 3]], [0, 1, 0], ["a", "b", "c"], [2, 1])

549 assert r.group_owners(0) == [0, 1, 2, 3]

550 assert r.group_owners(1) == [0, 1]

@@ -633,8 +614,7 @@ class TestReplicaOverlay:

614

615 def test_adding_a_deployment_does_not_move_the_first_one(self):

616 """The strongest form: an existing consumer's routes are byte-identical

636 - before and after another deployment is provisioned.

637 - """

617 + before and after another deployment is provisioned."""

618 one, names = self._fleet(16, 8, 1)

619 two, _ = self._fleet(16, 8, 2)

620 for consumer_id in range(8):

@@ -644,8 +624,7 @@ class TestReplicaOverlay:

624

625 def test_one_deployment_is_the_historical_carve(self):

626 """Default (no width) and width == the whole fleet must agree with the

647 - plain block rule, so a single-deployment fleet is untouched.

648 - """

627 + plain block rule, so a single-deployment fleet is untouched."""

628 plain = _router(16, 8, group_lens=[1] * 6)

629 explicit, names = self._fleet(16, 8, 1)

630 for consumer_id in range(8):

@@ -659,8 +638,7 @@ class TestReplicaOverlay:

638 def test_the_overlay_holds_for_partially_owned_names(self):

639 """Pipeline stages and experts too: two owner sets, one narrow. The

640 narrow one already forced agreement (one owner, no block to carve); the

662 - wide one is what the carve decides.

663 - """

641 + wide one is what the carve decides."""

642 r, names = self._fleet(

643 8,

644 4,

@@ -676,16 +654,14 @@ class TestReplicaOverlay:

654 def test_a_width_that_does_not_divide_the_fleet_raises(self):

655 """A uniform fleet is asserted upstream in ``get_world_size``; if it ever

656 is not, the overlay would map two workers of one deployment onto the same

679 - block index, so refuse rather than serve the wrong bytes.

680 - """

657 + block index, so refuse rather than serve the wrong bytes."""

658 with pytest.raises(ValueError, match="does not divide"):

659 RdtRouter(4, 6, None, None, ["a"], [1], workers_per_replica=4)

660

661 def test_pulls_still_carry_the_global_consumer_id(self):

662 """The carve uses the index WITHIN a deployment, but the wire must keep

663 the fleet-global id: it is what lets the producer tell the sharers of a

687 - slot apart, and count their arrivals separately.

688 - """

664 + slot apart, and count their arrivals separately."""

665

666 class _Method:

667 def __init__(self):

@@ -728,8 +704,7 @@ def _producer_pack_offsets(slices):

704 class TestPackedLayout:

705 def test_consumer_layout_matches_the_producer_rule(self):

706 """Mixed dtypes, sizes that do not land on 16B — i.e. the Kimi-style

731 - group (fp8 weights + fp32 scales + bf16 norms) that exposes alignment.

732 - """

707 + group (fp8 weights + fp32 scales + bf16 norms) that exposes alignment."""

708 specs = [

709 ("w.fp8", "float8_e4m3fn", [17]),

710 ("w.scale", "float32", [3]),

@@ -804,8 +779,7 @@ class TestPackedLayout:

779 class TestChunkModuleScatters:

780 """The chunk cut: one chunk per distinct owner class present in the copies,

781 ascending by class index. Derived purely from the bake plus the ownership

807 - table, so vLLM's expert placement needs no cases.

808 - """

782 + table, so vLLM's expert placement needs no cases."""

783

784 def test_unstamped_copies_form_one_chunk(self):

785 layer = _FakeLayer("l")

@@ -819,8 +793,7 @@ class TestChunkModuleScatters:

793 """Classes are numbered by first appearance in metadata order, so the

794 replicated class (0) leads and the rest follow — the ordering the old

795 "-1 first, then ascending stamp" rule produced, now falling out of the

822 - table instead of a special case.

823 - """

796 + table instead of a special case."""

797 layer = _FakeLayer("moe")

798 names = ["norm", "e5", "e0", "e2"]

799 copies = [_copy(n, shape=(4,)) for n in names]

@@ -840,8 +813,7 @@ class TestChunkModuleScatters:

813

814 def test_copy_order_within_a_chunk_is_bake_order(self):

815 """Scatter order within a chunk must follow the bake, not the stamp

843 - sort: the packed layout and the producer's replay agree on keys order.

844 - """

816 + sort: the packed layout and the producer's replay agree on keys order."""

817 layer = _FakeLayer("moe")

818 names = ["e3", "e1", "e2"] # one stamp, deliberately unsorted names

819 copies = [_copy(n, shape=(4,)) for n in names]

@@ -855,8 +827,7 @@ class TestChunkModuleScatters:

827

828 def test_stamps_interleave_across_modules_without_reordering_within(self):

829 """Two modules' copies bucket by stamp; within each bucket the order is

858 - module-then-bake order.

859 - """

830 + module-then-bake order."""

831 a, b = _FakeLayer("a"), _FakeLayer("b")

832 mods = [

833 _module(a, [_copy("a0", shape=(4,)), _copy("a1", shape=(4,))]),

@@ -876,8 +847,7 @@ class TestChunkModuleScatters:

847 def test_scatters_carry_their_own_dtype(self):

848 """Dtype rides the record from the bake, where it is the fake's dtype

849 AFTER its op chain — not a plan-time lookup of the source name, which

879 - would be wrong for any chain that reinterprets dtype.

880 - """

850 + would be wrong for any chain that reinterprets dtype."""

851 layer = _FakeLayer("l")

852 eng = _planner({}, name_meta={"w": ("float32", [10])})

853 ((_er, scatters),) = eng._chunk_module_scatters(

@@ -916,8 +886,7 @@ class TestBuildCallPlan:

886

887 def test_materialize_fires_on_a_modules_first_chunk_only(self):

888 """Empty HF params are allocated once per module, by construction —

919 - including a FusedMoE-like module whose copies span owner-class chunks.

920 - """

889 + including a FusedMoE-like module whose copies span owner-class chunks."""

890 layer = _FakeLayer("spanning")

891 names = [f"w{i}" for i in range(4)]

892 copies = [_copy(n, shape=(4,)) for n in names]

@@ -952,8 +921,7 @@ class TestBuildCallPlan:

921 def test_free_signal_waits_for_a_groups_last_chunk_when_it_is_split(self):

922 """The load-bearing case: with the group cut across owner-class chunks the

923 signal must hang off the LAST one. Signaling earlier lets the producers

955 - drop the gather buffers while a later chunk's RDMA is still reading.

956 - """

924 + drop the gather buffers while a later chunk's RDMA is still reading."""

925 layer = _FakeLayer("l")

926 names = [f"w{i}" for i in range(4)]

927 copies = [_copy(n, shape=(4,)) for n in names]

@@ -969,8 +937,7 @@ class TestBuildCallPlan:

937

938 def test_free_signal_timing_across_two_split_groups(self):

939 """Two groups x 2 owner-class chunks each: signals land on each group's

972 - last chunk only.

973 - """

940 + last chunk only."""

941 names_a = [f"model.layers.0.w{i}" for i in range(4)]

942 names_b = [f"model.layers.1.w{i}" for i in range(4)]

943 baked, name_meta, stamps = {}, {}, {}

@@ -988,8 +955,7 @@ class TestBuildCallPlan:

955 def test_a_group_with_nothing_local_is_signaled_at_sync_start(self):

956 """The owners still published it, so this consumer must signal it —

957 immediately, before the pipeline (signal-before-publish is tolerated),

991 - never hung off another group's chunk.

992 - """

958 + never hung off another group's chunk."""

959 baked, name_meta, names, group_lens = _one_module_per_layer(2)

960 del baked["model.layers.1.w"] # nothing baked, and not live => no pull

961 eng = _planner(

@@ -1010,8 +976,7 @@ class TestBuildCallPlan:

976

977 def test_modules_are_deduped_within_a_group(self):

978 """Several names of one fused module map to the same scatter list; the

1013 - plan must not chunk it twice.

1014 - """

979 + plan must not chunk it twice."""

980 layer = _FakeLayer("fused")

981 names = ["qkv.q", "qkv.k", "qkv.v"]

982 copies = [_copy(n, shape=(4,)) for n in names]

@@ -1027,8 +992,7 @@ class TestBuildCallPlan:

992

993 def test_plan_is_a_pure_function_of_its_inputs(self):

994 """Called twice with the same arguments it must produce the same plan —

1030 - that is what lets the engine build it once and reuse it every sync.

1031 - """

995 + that is what lets the engine build it once and reuse it every sync."""

996 baked, name_meta, names, group_lens = _one_module_per_layer(2)

997 eng = _planner(baked, name_meta=name_meta)

998 a = eng._build_call_plan(names, group_lens)

@@ -1048,8 +1012,7 @@ class TestBuildCallPlan:

1012 class TestUnbakedNamesGuard:

1013 """A live name with no baked plan has no way to load — there is no fallback,

1014 and its pull would target groups the pipeline already freed — so the plan

1051 - build fails at init, naming the names.

1052 - """

1015 + build fails at init, naming the names."""

1016

1017 def test_a_live_unbaked_name_fails_the_plan_build(self):

1018 baked, name_meta, names, group_lens = _one_module_per_layer(2)

@@ -1060,8 +1023,7 @@ class TestUnbakedNamesGuard:

1023

1024 def test_never_copied_names_are_dropped_entirely(self):

1025 """Experts owned by another EP rank no-op in their loader: not an error,

1063 - and not a chunk — their group is pre-freed if nothing else fills it.

1064 - """

1026 + and not a chunk — their group is pre-freed if nothing else fills it."""

1027 baked, name_meta, names, group_lens = _one_module_per_layer(2)

1028 del baked["norm.weight"]

1029 eng = _planner(baked, name_meta=name_meta, live=set(names) - {"norm.weight"})

@@ -1073,15 +1035,13 @@ class TestUnbakedNamesGuard:

1035 class TestCallPlanRouting:

1036 """Under partial ownership each chunk must be pulled from a producer that

1037 actually holds every name in it. Consumers bind EVERY producer (signals fan

1076 - out to all owners), so local index == trainer rank.

1077 - """

1038 + out to all owners), so local index == trainer rank."""

1039

1040 def _routed_planner(

1041 self, held_by_group, num_consumers, consumer_id, *, held_by=None

1042 ):

1043 """``held_by_group[g]`` = the ranks holding group g's names, so a test

1083 - can still speak in groups; ``held_by`` overrides individual names.

1084 - """

1044 + can still speak in groups; ``held_by`` overrides individual names."""

1045 baked, name_meta, names, group_lens = _one_module_per_layer(

1046 len(held_by_group) - 2

1047 )

@@ -1108,16 +1068,14 @@ class TestCallPlanRouting:

1068

1069 def test_two_stage_ownership_splits_chunks_between_owners(self):

1070 """Two PP stages, 4 groups: stage 0 holds the first half, stage 1 the

1111 - second. Every chunk goes to a holder of its names.

1112 - """

1071 + second. Every chunk goes to a holder of its names."""

1072 eng, names, group_lens = self._routed_planner([[0], [0], [1], [1]], 1, 0)

1073 plan = eng._build_call_plan(names, group_lens)

1074 assert [c.owner for c in plan.chunks] == [0, 0, 1, 1]

1075

1076 def test_a_multi_owner_group_still_resolves_to_one_producer(self):

1077 """Every chunk is served by exactly ONE producer: splitting a pull only

1119 - multiplies produce calls, since the consumer's NIC bounds it.

1120 - """

1078 + multiplies produce calls, since the consumer's NIC bounds it."""

1079 eng, names, group_lens = self._routed_planner([[0, 1]] * 4, 1, 0)

1080 plan = eng._build_call_plan(names, group_lens)

1081 assert all(0 <= c.owner < 2 for c in plan.chunks)

@@ -1125,8 +1083,7 @@ class TestCallPlanRouting:

1083 def test_an_expert_name_routes_to_the_rank_holding_it(self):

1084 """PP ∩ EP: stage 0 = ranks 0-1, stage 1 = ranks 2-3, and one expert

1085 name inside a stage-1 group is held by rank 3 alone. The name's own owner

1128 - set is what routes it.

1129 - """

1086 + set is what routes it."""

1087 eng, names, group_lens = self._routed_planner(

1088 [[0, 1], [0, 1], [2, 3], [2, 3]],

1089 1,

@@ -1141,8 +1098,7 @@ class TestCallPlanRouting:

1098 def test_an_unheld_name_is_rejected_at_init(self):

1099 """A name no rank holds can never be served. The trainer raises when it

1100 transposes the holdings; the router's validate() is the consumer-side

1144 - backstop, and either way it fails at init rather than at first pull.

1145 - """

1101 + backstop, and either way it fails at init rather than at first pull."""

1102 with pytest.raises(ValueError, match="empty"):

1103 RdtRouter(2, 1, [[]], [0], ["w"], [1]).validate()

1104

@@ -1151,16 +1107,14 @@ class TestSignalCompleteness:

1107 """Every gather group is signaled exactly once — after its last chunk when

1108 the worker pulls from it, at sync start otherwise — for ANY expert

1109 placement. A group signaled twice over-credits the barrier; a group never

1154 - signaled parks a producer credit and hangs end_sync.

1155 - """

1110 + signaled parks a producer credit and hangs end_sync."""

1111

1112 def _moe_planner(self, worker_experts, *, ep_size=4, n_experts=8, n_layers=2):

1113 """Pre / n_layers MoE layers / post. Each layer: one norm (its own

1114 module, stamp -1) + n_experts expert names stamped ``e // n_local``.

1115 The worker's bake covers only ``worker_experts`` (its placement);

1116 foreign expert names never copied => dropped entirely, as in the real

1162 - bake.

1163 - """

1117 + bake."""

1118 n_local = n_experts // ep_size

1119 names = ["embed.weight"]

1120 group_lens = [1]

@@ -1220,8 +1174,7 @@ class TestSignalCompleteness:

1174

1175 def test_placement_changes_only_the_chunk_count_never_the_signals(self):

1176 """Linear experts 0-3 hit 2 owner classes; round_robin 0,2,4,6

1223 - hits all 4. More chunks per group, identical signal set.

1224 - """

1177 + hits all 4. More chunks per group, identical signal set."""

1178 plans = {}

1179 for label, experts in (("linear", [0, 1, 2, 3]), ("round_robin", [0, 2, 4, 6])):

1180 eng, names, group_lens = self._moe_planner(worker_experts=experts)

@@ -1235,8 +1188,7 @@ class TestSignalCompleteness:

1188

1189 def test_the_fused_module_materializes_first_and_quants_last_across_chunks(self):

1190 """The FusedMoE shape: one module's copies span every owner-class chunk of

1238 - its group; materialize on its first, quant on its last.

1239 - """

1191 + its group; materialize on its first, quant on its last."""

1192 eng, names, group_lens = self._moe_planner(

1193 worker_experts=[0, 2, 4, 6], n_layers=1

1194 )

@@ -1294,8 +1246,7 @@ class TestLayerwiseGroups:

1246 def test_flattening_a_partition_returns_the_input_order(self):

1247 """Un-indexed names group by POSITION, not by name class — so one that

1248 looks like a pre-block name still lands after the layers it follows.

1297 - Group index therefore means the same thing on every rank.

1298 - """

1249 + Group index therefore means the same thing on every rank."""

1250 names = [

1251 "embed.weight",

1252 "model.layers.0.a",

@@ -1313,8 +1264,7 @@ class TestLayerwiseGroups:

1264

1265 def test_stacks_keep_prefix_appearance_order_and_sort_within(self):

1266 """Two stacks stay in the order they appear — the vision tower before the

1316 - text stack — while each sorts by index internally.

1317 - """

1267 + text stack — while each sorts by index internally."""

1268 names = [

1269 "visual.blocks.1.w",

1270 "visual.blocks.0.w",

@@ -1331,8 +1281,7 @@ class TestLayerwiseGroups:

1281 def test_the_post_block_lands_last_however_early_it_arrives(self):

1282 """Megatron-Bridge streams the last pipeline stage's output block before

1283 its layers. Sweeping un-indexed names after the first layer into a

1334 - trailing group is what lets the gather loop walk groups ascending.

1335 - """

1284 + trailing group is what lets the gather loop walk groups ascending."""

1285 names = ["model.layers.0.w", "model.norm.weight", "model.layers.1.w"]

1286 assert layerwise_groups(names) == [

1287 ["model.layers.0.w"],

@@ -1342,8 +1291,7 @@ class TestLayerwiseGroups:

1291

1292 def test_names_sharing_a_layer_coalesce_when_the_source_interleaves(self):

1293 """A raw-checkpoint source may yield in shard-packing order. Coalescing on

1345 - the key keeps that from shattering one layer into many groups.

1346 - """

1294 + the key keeps that from shattering one layer into many groups."""

1295 names = [

1296 "model.layers.0.a",

1297 "model.layers.1.a",

@@ -1385,15 +1333,13 @@ class TestLayerwiseGroups:

1333 def test_every_supported_naming_convention_partitions_per_layer(self, prefix):

1334 """One decoder layer per group under any prefix, with no per-architecture

1335 table. A literal `model.layers.` match silently yields one whole-model

1388 - group for most of these.

1389 - """

1336 + group for most of these."""

1337 names = [f"{prefix}{i}.self_attn.q_proj.weight" for i in range(3)]

1338 assert layerwise_groups(names) == [[n] for n in names]

1339

1340 def test_a_moe_layer_stays_whole(self):

1341 """The OUTERMOST index wins, so per-expert names group by layer. Splitting

1395 - on the expert would cut FusedMoE — one leaf module — across groups.

1396 - """

1342 + on the expert would cut FusedMoE — one leaf module — across groups."""

1343 names = [

1344 "model.layers.0.mlp.experts.0.w1.weight",

1345 "model.layers.0.mlp.experts.1.w1.weight",

@@ -1410,8 +1356,7 @@ class TestLayerwiseGroups:

1356 def test_a_vlm_partitions_the_vision_tower_and_the_text_stack(self):

1357 """Qwen-VL layout. Today's literal `model.layers.` match puts every one of

1358 these in ONE group — the whole model — because the text stack ships as

1413 - `model.language_model.layers.`.

1414 - """

1359 + `model.language_model.layers.`."""

1360 names = [

1361 "visual.patch_embed.proj.weight",

1362 "visual.blocks.0.attn.qkv.weight",

@@ -1430,8 +1375,7 @@ class TestLayerwiseGroups:

1375

1376 def test_two_stacks_sharing_an_index_do_not_merge(self):

1377 """The key carries the prefix, so vision block 0 and text layer 0 are

1433 - different groups even though both are index 0.

1434 - """

1378 + different groups even though both are index 0."""

1379 names = [

1380 "visual.blocks.0.attn.qkv.weight",

1381 "model.language_model.layers.0.self_attn.q_proj.weight",

@@ -1468,23 +1412,20 @@ class TestOpAllowlistAgreement:

1412

1413 def test_every_bake_recordable_op_is_replayable(self):

1414 """Anything the bake can record must be serveable, or the failure lands a

1471 - whole sync later than the mistake.

1472 - """

1415 + whole sync later than the mistake."""

1416 assert not set(SUPPORTED_OPS.values()) - set(ALLOWED_OPS)

1417

1418 def test_the_producer_allows_nothing_the_consumer_cannot_emit(self):

1419 """The allowlist is also a guard against a misbehaving or spoofed

1420 consumer invoking arbitrary methods on trainer tensors, so it must not be

1421 wider than the ops the bake can actually produce. ``to`` in particular

1479 - would let a replay change dtype or device.

1480 - """

1422 + would let a replay change dtype or device."""

1423 assert not set(ALLOWED_OPS) - set(SUPPORTED_OPS.values())

1424 assert "to" not in ALLOWED_OPS

1425

1426 def test_transpose_via_t_is_serveable(self):

1427 """The specific regression: a chain recorded from ``.t()`` must pass the

1486 - producer's guard.

1487 - """

1428 + producer's guard."""

1429 fake = FakeRDTTensor(

1430 name="w",

1431 shape=torch.Size((4, 6)),

@@ -1529,8 +1470,7 @@ BAKE_MODELS = [

1470 class TestRequiresTheRayExecutor:

1471 """The engine's data plane is Ray's, so a non-Ray executor cannot work. The

1472 check is at construction because the alternative is an opaque failure during

1532 - the first handshake, long after the misconfiguration.

1533 - """

1473 + the first handshake, long after the misconfiguration."""

1474

1475 def _construct(self, backend):

1476 cfg = SimpleNamespace(

@@ -1557,8 +1497,7 @@ class TestRequiresTheRayExecutor:

1497

1498 def test_a_custom_executor_class_is_left_alone(self):

1499 """A `type[Executor]` override is deliberate and unjudgeable here, so it

1560 - must not be rejected for merely not being the string "ray".

1561 - """

1500 + must not be rejected for merely not being the string "ray"."""

1501

1502 class _CustomExecutor:

1503 pass

tests/distributed/test_sharded_rdt_producer.py

+39 -78

@@ -236,8 +236,7 @@ class TestPublishAndRebuild:

236 def test_views_of_one_storage_do_not_overlap(self, server_factory):

237 """One IPC export per storage, one as_strided view per name — the whole

238 point of the storage/view split. Overlapping views would serve the same

239 - bytes for two different weights.

240 - """

239 + bytes for two different weights."""

240 server = server_factory()

241 server.begin_sync(1)

242 _publish(server, GI_A, GROUP_A, nbytes=64)

@@ -260,8 +259,7 @@ class TestPublishAndRebuild:

259 def test_the_default_lookahead_is_one(self):

260 """1 is the sweet spot: group N+1 is gathered AND pullable while N is

261 being pulled, with resident memory at its floor of 2 groups — the bound

263 - the larger-model runs size against.

264 - """

262 + the larger-model runs size against."""

263 assert DEFAULT_GATHER_LOOKAHEAD == 1

264

265

@@ -276,8 +274,7 @@ class TestFreeBarrier:

274

275 def test_group_is_held_until_the_last_live_consumer_signals(self, server_factory):

276 """Every live consumer signals every owner; releasing on the first

279 - signal would drop storage other consumers are still reading.

280 - """

277 + signal would drop storage other consumers are still reading."""

278 server = server_factory()

279 server.begin_sync(3)

280 _publish(server, GI_A, GROUP_A)

@@ -304,8 +301,7 @@ class TestFreeBarrier:

301 ):

302 """A consumer with nothing to pull from a group signals it at sync

303 start, which can precede the publish. The publish must then release the

307 - group rather than wait for a signal that will never come again.

308 - """

304 + group rather than wait for a signal that will never come again."""

305 server = server_factory()

306 server.begin_sync(1)

307 server.free_group(GI_A)

@@ -328,8 +324,7 @@ class TestFreeBarrier:

324 the server says the group is gone — and it hears that ONLY through

325 ``wait_freed`` / ``end_sync``, never a publish return (a freed notice

326 riding an unharvested async publish while the engine blocks in

331 - ``wait_freed`` would wedge the loop).

332 - """

327 + ``wait_freed`` would wedge the loop)."""

328 server = server_factory()

329 server.begin_sync(1)

330 _publish(server, GI_A, GROUP_A)

@@ -358,8 +353,7 @@ class TestFreeBarrier:

353

354 class _CountingSource:

355 """`iter_groups` yielding one CPU-tensor group per call, built lazily so a

361 - group's tensors exist only once the engine's credit gate lets it gather.

362 - """

356 + group's tensors exist only once the engine's credit gate lets it gather."""

357

358 def __init__(self, groups):

359 self._groups = groups

@@ -372,8 +366,7 @@ class _CountingSource:

366 class _ResidencyDict(dict):

367 """The engine's `_inflight`, instrumented: `max_len` is the high-water mark

368 of groups whose CUDA-IPC export refs were alive at once — the trainer's

375 - resident-memory bound in groups.

376 - """

369 + resident-memory bound in groups."""

370

371 def __init__(self):

372 super().__init__()

@@ -387,8 +380,7 @@ class _ResidencyDict(dict):

380 def _loop_engine(server, n_groups, *, lookahead):

381 """A real engine wired straight to a real (non-Ray) server, with only the

382 state `_run_gather_loop` reads. The `_rpc` seam dispatches inline;

390 - `_publish_async` sees no `.remote` and runs inline too.

391 - """

383 + `_publish_async` sees no `.remote` and runs inline too."""

384 groups = [(f"model.layers.{gi}.w",) for gi in range(n_groups)]

385 e = ShardedRDTTrainerWeightTransferEngine.__new__(

386 ShardedRDTTrainerWeightTransferEngine

@@ -411,16 +403,14 @@ class TestGatherCredit:

403 serveable immediately, and the ENGINE's loop stops gathering while more than

404 `gather_lookahead` groups are unfreed. So at most `lookahead + 1` groups are

405 resident — two at the default — while group N+1 is pulled the instant N's

414 - pulls finish.

415 - """

406 + pulls finish."""

407

408 @pytest.fixture

409 def gather_engine(self, monkeypatch):

410 """Neutralize the two CUDA touches in the export path so the REAL

411 gather loop runs on CPU: `.cuda()` becomes identity, and the storage

412 export ships the `(nbytes, ..., device, ...)` tuple the fixture's

422 - stubbed `rebuild_cuda_tensor` expects.

423 - """

413 + stubbed `rebuild_cuda_tensor` expects."""

414 monkeypatch.setattr(torch.Tensor, "cuda", lambda self: self)

415 monkeypatch.setattr(

416 trainer_mod,

@@ -432,8 +422,7 @@ class TestGatherCredit:

422 def test_publish_never_blocks(self, server_factory):

423 """Publishing must not park on a credit: the whole point is that a

424 gathered group becomes serveable immediately. Three publishes with no

435 - frees, inline on this thread — a block here hangs the test.

436 - """

425 + frees, inline on this thread — a block here hangs the test."""

426 server = server_factory(gather_lookahead=1)

427 server.begin_sync(1)

428 for gi in range(3):

@@ -447,8 +436,7 @@ class TestGatherCredit:

436 than 2 groups of gathered weights, no matter how the consumer paces —

437 and the pipeline still overlaps, because the consumer here refuses to

438 free group N until N+1 is already published (pullable). A gate an

450 - off-by-one too tight deadlocks this test; too loose fails max_len.

451 - """

439 + off-by-one too tight deadlocks this test; too loose fails max_len."""

440 server = server_factory(gather_lookahead=1)

441 n_groups = 8

442 engine = gather_engine(server, n_groups, lookahead=1)

@@ -490,8 +478,7 @@ class TestGatherCredit:

478 ):

479 """0 is the serialized baseline: the loop waits for group N to be freed

480 before gathering N+1, so exactly ONE group is ever resident and nothing

493 - overlaps. A floor of 1 on the bound would silently pipeline instead.

494 - """

481 + overlaps. A floor of 1 on the bound would silently pipeline instead."""

482 server = server_factory(gather_lookahead=0)

483 engine = gather_engine(server, 4, lookahead=0)

484

@@ -517,8 +504,7 @@ class TestGatherCredit:

504 self, server_factory, gather_engine

505 ):

506 """Lookahead + 1, not a hardcoded 2: at lookahead=2 a free-nothing

520 - consumer sees exactly 3 groups gathered before the loop parks.

521 - """

507 + consumer sees exactly 3 groups gathered before the loop parks."""

508 server = server_factory(gather_lookahead=2)

509 engine = gather_engine(server, 8, lookahead=2)

510

@@ -554,8 +540,7 @@ class TestGatherCredit:

540 def test_a_gather_error_releases_a_blocked_wait_freed(self, server_factory):

541 """Otherwise a trainer-side failure on another rank deadlocks this one

542 inside its credit gate. It must RAISE, not return empty — an empty

557 - return would spin the engine straight back into the wait.

558 - """

543 + return would spin the engine straight back into the wait."""

544 server = server_factory()

545 server.begin_sync(1)

546 _publish(server, GI_A, GROUP_A)

@@ -636,8 +621,7 @@ class TestBeginSync:

621

622 def test_the_target_is_floored_at_one(self, server_factory):

623 """A bare/zero call must not make every group free instantly (or divide

639 - the barrier by zero).

640 - """

624 + the barrier by zero)."""

625 server = server_factory()

626 server.begin_sync(0)

627 assert server._live_count == 1

@@ -645,8 +629,7 @@ class TestBeginSync:

629 def test_a_degraded_sync_lowers_the_target(self, server_factory):

630 """FT degraded sync: the live count is the WHOLE degraded-sync

631 mechanism on this side — the group releases after the live consumers

648 - alone.

649 - """

632 + alone."""

633 server = server_factory()

634 server.begin_sync(2) # 4 provisioned, 2 alive

635 _publish(server, GI_A, GROUP_A)

@@ -670,8 +653,7 @@ class TestBeginSync:

653 def test_a_straggler_signal_cannot_credit_the_next_sync(self, server_factory):

654 """begin_sync resets the counts, so a signal that leaked past the sync

655 boundary would over-credit — the reason the consumer drains its fired

673 - signals before finishing. Pin the reset side of that contract.

674 - """

656 + signals before finishing. Pin the reset side of that contract."""

657 server = server_factory()

658 server.begin_sync(2)

659 _publish(server, GI_A, GROUP_A)

@@ -693,8 +675,7 @@ class TestBeginSync:

675 self, server_factory

676 ):

677 """The packed layout repeats every sync — that is what makes caching the

696 - destination views worth ~5ms per 384-spec group.

697 - """

678 + destination views worth ~5ms per 384-spec group."""

679 server = server_factory()

680 server._pack_dsts[("sentinel",)] = (0, [])

681 server.begin_sync(1)

@@ -705,8 +686,7 @@ class TestServedNamesGuard:

686 """A producer serves only the names it holds: its stage's groups, and

687 within them its own EP coordinate's experts plus the replicated names.

688 Without this guard a misrouted pull would block forever in the cache wait

708 - for a name this rank never publishes.

709 - """

689 + for a name this rank never publishes."""

690

691 def test_a_pull_for_an_unserved_name_fails_loudly(self, server_factory):

692 server = server_factory(served_names=list(GROUP_A))

@@ -722,8 +702,7 @@ class TestServedNamesGuard:

702

703 def test_served_names_none_accepts_anything(self, server_factory):

704 """Gather-to-all: every producer holds every group, so there is nothing

725 - to guard. The call proceeds to the cache wait (not exercised here).

726 - """

705 + to guard. The call proceeds to the cache wait (not exercised here)."""

706 server = server_factory(served_names=None)

707 assert server._served_names is None

708

@@ -770,8 +749,7 @@ class TestStallWatchdog:

749

750 def test_end_sync_fails_when_a_consumer_never_signals(self, server_factory):

751 """The exact mid-sync death signature: the group is published, one of

773 - the two live consumers that owed a signal is gone.

774 - """

752 + the two live consumers that owed a signal is gone."""

753 server = server_factory(stall_timeout_s=0.3)

754 server.begin_sync(2)

755 _publish(server, GI_A, GROUP_A)

@@ -785,8 +763,7 @@ class TestStallWatchdog:

763 """The watchdog fires on the existing `set_gather_error` channel, so one

764 stall unwinds the whole rank through one path rather than each waiter

765 timing out separately — here both a consumer parked in the serve cache

788 - wait and the engine parked in its credit gate.

789 - """

766 + wait and the engine parked in its credit gate."""

767 server = server_factory(gather_lookahead=1, stall_timeout_s=0.3)

768 server.begin_sync(1)

769 _publish(server, GI_A, GROUP_A)

@@ -819,8 +796,7 @@ class TestStallWatchdog:

796 """A consumer that is slow but signaling must never trip it: the stamp

797 is global to the producer, so a steady trickle of signals holds the

798 whole rank open. Nine sequential publishes with a 0.3s timeout each take

822 - longer in total than the timeout, and none may fire.

823 - """

799 + longer in total than the timeout, and none may fire."""

800 server = server_factory(gather_lookahead=1, stall_timeout_s=0.3)

801 server.begin_sync(1)

802 for gi in range(9):

@@ -833,8 +809,7 @@ class TestStallWatchdog:

809

810 def test_begin_sync_resets_the_progress_stamp(self, server_factory):

811 """Syncs are minutes apart. Without the reset the first publish of sync N+1

836 - would measure its stall from somewhere inside sync N and fire immediately.

837 - """

812 + would measure its stall from somewhere inside sync N and fire immediately."""

813 server = server_factory(gather_lookahead=1, stall_timeout_s=0.3)

814 server.begin_sync(1)

815 _publish(server, GI_A, GROUP_A)

@@ -856,8 +831,7 @@ class TestConcurrentPublishAndFree:

831 self, server_factory

832 ):

833 """Every group published must end up freed and out of the cache, with

859 - end_sync returning cleanly — the property whose violation is a hang.

860 - """

834 + end_sync returning cleanly — the property whose violation is a hang."""

835 server = server_factory(gather_lookahead=2)

836 server.begin_sync(1)

837 groups = [(gi, (f"model.layers.{gi}.w",)) for gi in range(12)]

@@ -892,8 +866,7 @@ class TestConcurrentPublishAndFree:

866 ):

867 """All live consumers signaling at once: the release must happen on the

868 last signal and only once, or the engine drops IPC refs twice (or

895 - never).

896 - """

869 + never)."""

870 server = server_factory()

871 server.begin_sync(8)

872 _publish(server, GI_A, GROUP_A)

@@ -920,8 +893,7 @@ class TestFakeServerAgreesWithTheRealOne:

893 """`test_sharded_rdt_trainer.py`'s `_FakeProducerServer` is a second,

894 independent model of this protocol that the engine-side tests assert on. Pin

895 the two against each other so the fake cannot silently drift from the

923 - semantics the real server enforces.

924 - """

896 + semantics the real server enforces."""

897

898 @staticmethod

899 def _fake():

@@ -952,8 +924,7 @@ class TestExportRing:

924 """The packed export path. Its one hard invariant is that the trainer packs

925 a group exactly the way ``publish_group`` rebuilds it — a layout

926 disagreement would serve the wrong bytes with nothing downstream to catch

955 - it.

956 - """

927 + it."""

928

929 @staticmethod

930 def _engine():

@@ -1045,8 +1016,7 @@ class TestExportRingSlotSafety:

1016 ):

1017 """Free the later group of each pair first and hold the earlier one live

1018 until the NEXT group is gathered -- the window a modular slot counter got

1048 - wrong (at lookahead=1 there are 2 slots, so group 2 took group 0's).

1049 - """

1019 + wrong (at lookahead=1 there are 2 slots, so group 2 took group 0's)."""

1020 server = server_factory(gather_lookahead=1)

1021 n_groups = 8

1022 engine = gather_engine(server, n_groups, lookahead=1)

@@ -1170,8 +1140,7 @@ class TestSharedServeSlots:

1140

1141 def test_one_deployment_never_shares(self, sharing_server):

1142 """With the width equal to the live count every group is a singleton, so

1173 - each call packs for itself out of its own ring.

1174 - """

1143 + each call packs for itself out of its own ring."""

1144 server = sharing_server(workers_per_replica=2)

1145 server.begin_sync(2, [0, 1])

1146 _publish(server, GI_A, GROUP_A)

@@ -1251,8 +1220,7 @@ class TestSharedServeSlots:

1220

1221 def test_a_seqless_call_is_refused(self, sharing_server):

1222 """The slot comes from ``seq``, so a caller that sends none cannot be

1254 - served safely.

1255 - """

1223 + served safely."""

1224 server = sharing_server(workers_per_replica=2)

1225 server.begin_sync(1, [0])

1226 _publish(server, GI_A, GROUP_A)

@@ -1262,8 +1230,7 @@ class TestSharedServeSlots:

1230

1231 def test_a_failed_pack_reaches_every_sharer(self, sharing_server):

1232 """Only the packer touches the buffer, so its failure has to be

1265 - republished: a waiter must raise rather than read a half-written slot.

1266 - """

1233 + republished: a waiter must raise rather than read a half-written slot."""

1234 server = sharing_server(workers_per_replica=2)

1235 server.begin_sync(4, [0, 1, 2, 3])

1236 _publish(server, GI_A, GROUP_A)

@@ -1287,8 +1254,7 @@ class TestSharedServeSlots:

1254

1255 def test_a_degraded_sync_narrows_the_rendezvous(self, sharing_server):

1256 """A dead deployment must not be waited for: a survivor left alone in

1290 - its group packs for itself instead of blocking on the corpse.

1291 - """

1257 + its group packs for itself instead of blocking on the corpse."""

1258 server = sharing_server(workers_per_replica=2, stall_timeout_s=30.0)

1259 server.begin_sync(3, [0, 1, 3]) # consumer 2 died

1260 _publish(server, GI_A, GROUP_A)

@@ -1302,8 +1268,7 @@ class TestSharedServeSlots:

1268

1269 def test_a_sharer_that_never_arrives_trips_the_watchdog(self, sharing_server):

1270 """A consumer that dies INSIDE the sync window has no detector, exactly

1305 - as for the free barrier; the rendezvous waits on the same channel.

1306 - """

1271 + as for the free barrier; the rendezvous waits on the same channel."""

1272 server = sharing_server(workers_per_replica=2, stall_timeout_s=0.3)

1273 server.begin_sync(4, [0, 1, 2, 3])

1274 _publish(server, GI_A, GROUP_A)

@@ -1315,8 +1280,7 @@ class TestSharedServeSlots:

1280

1281 class TestSharedReservation:

1282 """``reserve_serve_buffer`` under sharing: one ring per GROUP, and the

1318 - init-time check that the group's members pull the same chunks.

1319 - """

1283 + init-time check that the group's members pull the same chunks."""

1284

1285 def test_matching_plans_reserve_one_ring_between_them(self, sharing_server):

1286 server = sharing_server(workers_per_replica=2)

@@ -1338,8 +1302,7 @@ class TestSharedReservation:

1302 def test_mismatched_plans_fail_at_reservation(self, sharing_server):

1303 """Sharing rests on the sharers pulling the same chunks. Checked here,

1304 where it names both consumers, rather than at serve time, where it is a

1341 - rendezvous nobody completes.

1342 - """

1305 + rendezvous nobody completes."""

1306 server = sharing_server(workers_per_replica=2)

1307 server.reserve_serve_buffer(0, 1024, "plan-a")

1308

@@ -1348,8 +1311,7 @@ class TestSharedReservation:

1311

1312 def test_different_groups_may_differ(self, sharing_server):

1313 """Only the members of ONE group have to agree; worker 0 and worker 1 of

1351 - the same deployment hold different slices and never share a slot.

1352 - """

1314 + the same deployment hold different slices and never share a slot."""

1315 server = sharing_server(workers_per_replica=2)

1316 server.reserve_serve_buffer(0, 1024, "plan-a")

1317 server.reserve_serve_buffer(1, 1024, "plan-b")

@@ -1358,8 +1320,7 @@ class TestSharedReservation:

1320

1321 def test_a_missing_digest_skips_the_check(self, sharing_server):

1322 """The digest is optional on the wire, so an engine that does not send

1361 - one still reserves.

1362 - """

1323 + one still reserves."""

1324 server = sharing_server(workers_per_replica=2)

1325 server.reserve_serve_buffer(0, 1024, None)

1326 server.reserve_serve_buffer(2, 1024, None)

tests/distributed/test_sharded_rdt_trainer.py

+12 -24

@@ -34,8 +34,7 @@ from vllm.distributed.weight_transfer.sharded_rdt_trainer import (

34

35 class _ListSource(WeightSource):

36 """A WeightSource over an explicit ordered (name, cpu-tensor) list, so the

37 - sharded-RDT group/order logic can be tested without a real model.

38 - """

37 + sharded-RDT group/order logic can be tested without a real model."""

38

39 def __init__(self, pairs):

40 self._pairs = list(pairs)

@@ -55,8 +54,7 @@ class _FakeProducerServer:

54 per-group barrier: publish/free are keyed by GROUP INDEX, signals count

55 to the ``begin_sync`` live total, publish_group returns nothing, and freed

56 groups flow back only through wait_freed / end_sync (see

58 - test_sharded_rdt_producer.TestFakeServerAgreesWithTheRealOne).

59 - """

57 + test_sharded_rdt_producer.TestFakeServerAgreesWithTheRealOne)."""

58

59 def __init__(self, auto_free=True):

60 self.order: list[str] = []

@@ -100,8 +98,7 @@ class _FakeProducerServer:

98 must already have a banked credit when the gate asks (auto_free, or a

99 test's own free_group) — anything else is the deadlock the real

100 watchdog would kill, so fail loudly. Not appended to ``order``: it is

103 - pacing, not a lifecycle milestone.

104 - """

101 + pacing, not a lifecycle milestone."""

102 assert self._pending_freed, (

103 "wait_freed with nothing freed: the gather loop would deadlock"

104 )

@@ -124,8 +121,7 @@ def _rdt_engine_with_fake_server(

121 ):

122 """Build a ShardedRDTTrainerWeightTransferEngine wired to an in-process fake

123 server (no Ray, no CUDA IPC): bypass trainer_init's spawn, set the

127 - group-major metadata, and route _rpc to the fake.

128 - """

124 + group-major metadata, and route _rpc to the fake."""

125 import vllm.distributed.weight_transfer.sharded_rdt_trainer as mod

126

127 # reduce_tensor needs CUDA; the fake server never rebuilds, so stub it.

@@ -325,8 +321,7 @@ def test_sharded_rdt_send_weights_surfaces_update_error(monkeypatch):

321

322 class _OwnedSource(_ListSource):

323 """A source holding only some groups' names, like a pipeline-parallel rank.

328 - Takes group indices for convenience and declares the names inside them.

329 - """

324 + Takes group indices for convenience and declares the names inside them."""

325

326 def __init__(self, pairs, owned_group_idx):

327 super().__init__(pairs)

@@ -392,8 +387,7 @@ def test_sharded_rdt_owned_group_order_mismatch_raises(monkeypatch):

387 )

388 def test_sharded_rdt_begin_sync_carries_the_live_count(monkeypatch):

389 """The free barrier's target is one integer per sync: the live consumer

395 - count, defaulting to the whole provisioned fleet.

396 - """

390 + count, defaulting to the whole provisioned fleet."""

391 server = _FakeProducerServer(auto_free=True)

392 engine = _rdt_engine_with_fake_server(

393 _rdt_source_two_layers(),

@@ -411,8 +405,7 @@ def test_sharded_rdt_begin_sync_carries_the_live_count(monkeypatch):

405 class TestLiveCountPlumbing:

406 """``send_weights(live_consumer_ids)`` -> the barrier target. The

407 provisioned geometry is frozen; a degraded sync only lowers the target and

414 - narrows each slot-sharing group's rendezvous.

415 - """

408 + narrows each slot-sharing group's rendezvous."""

409

410 @staticmethod

411 def _engine(num_consumers, world=2, rank=0):

@@ -445,8 +438,7 @@ class TestLiveCountPlumbing:

438

439 def test_which_consumers_died_does_not_matter_only_how_many(self):

440 """The whole point of the barrier: no routed per-producer targets, so

448 - the identity of the dead consumer is irrelevant to the producers.

449 - """

441 + the identity of the dead consumer is irrelevant to the producers."""

442 counts = []

443 for live in ([0, 1, 2, 3], [4, 5, 6, 7], [0, 2, 4, 6]):

444 engine, got = self._engine(num_consumers=8)

@@ -457,8 +449,7 @@ class TestLiveCountPlumbing:

449 def test_the_live_ids_travel_with_the_count(self):

450 """The count sizes the free barrier, the ids size the slot-sharing

451 rendezvous, so a producer that shares slots can tell WHICH consumers it

460 - is still waiting for. They must describe the same set.

461 - """

452 + is still waiting for. They must describe the same set."""

453 engine, got = self._engine(num_consumers=8)

454 engine.send_weights([6, 0, 2, 2])

455 ((count, ids),) = got

@@ -526,8 +517,7 @@ def test_sharded_rdt_rejects_a_rank_holding_nothing(monkeypatch):

517

518 def test_sharded_rdt_rejects_a_name_no_rank_holds(monkeypatch):

519 """Every name must be held somewhere or it can never be served — caught

529 - when the holdings are transposed, naming the orphan.

530 - """

520 + when the holdings are transposed, naming the orphan."""

521 with pytest.raises(ValueError, match="no trainer rank holds"):

522 _rdt_engine_with_fake_server(

523 _OwnedSource(_rdt_source_two_layers()._pairs, [0, 1]),

@@ -541,8 +531,7 @@ def test_sharded_rdt_rejects_a_name_no_rank_holds(monkeypatch):

531

532 def test_sharded_rdt_rejects_metadata_disagreement_across_ranks(monkeypatch):

533 """Only the sender's metadata reaches the consumers, so a rank describing

544 - just its own share must fail loudly rather than silently drop the rest.

545 - """

534 + just its own share must fail loudly rather than silently drop the rest."""

535 engine = _rdt_engine_with_fake_server(

536 _OwnedSource(_rdt_source_two_layers()._pairs, [1, 2]),

537 is_sender=False,

@@ -565,8 +554,7 @@ def _serve_ring_server(src_name, src):

554 """A producer server with one cached tensor and a pre-seeded serve ring, so a

555 pull needs no Ray and no NIXL registration. Returns (server, serve) where

556 ``serve(chain)`` packs one spec into the SAME ring slot every time — which is

568 - what puts the destination-view cache, and only it, under test.

569 - """

557 + what puts the destination-view cache, and only it, under test."""

558 from vllm.distributed.weight_transfer.sharded_rdt_trainer import (

559 _RDTProducerServer,

560 )

tests/distributed/test_shm_broadcast.py

+3 -6

@@ -419,8 +419,7 @@ def test_tensor_broadcast():

419

420 def _dumps_oob(obj) -> tuple[bytes, list]:

421 """Pickle `obj` the same way `MessageQueue.enqueue` does: tensor

422 - dispatch table + out-of-band buffers >= 1MiB.

423 - """

422 + dispatch table + out-of-band buffers >= 1MiB."""

423 buffers = []

424

425 def callback(buf: pickle.PickleBuffer) -> bool:

@@ -491,8 +490,7 @@ def test_tensor_pickle_roundtrip(case: str):

490 @pytest.mark.parametrize("case", ["cuda", "requires_grad", "conj"])

491 def test_reduce_tensor_fallback(case: str):

492 """Tensors the zero-copy reducer can't safely alias must fall back to

494 - torch's default reduction.

495 - """

493 + torch's default reduction."""

494 if case == "cuda":

495 if not torch.cuda.is_available():

496 pytest.skip("requires CUDA")

@@ -755,8 +753,7 @@ def test_remote_subscribe_addr_unique_concurrent_writers(

753 Pre-fix, the writer probed a port with get_open_port() and bound it

754 afterwards; pinning the probe to one free port makes every writer

755 bind the same port and fail deterministically on that code path,

758 - while the late-binding implementation never consults the probe.

759 - """

756 + while the late-binding implementation never consults the probe."""

757 from vllm.distributed.device_communicators import shm_broadcast

758

759 colliding_port = get_open_port()

tests/distributed/test_utils.py

+1 -2

@@ -31,8 +31,7 @@ class _CUDADeviceCountStatelessTestActor:

31

32 def test_cuda_device_count_stateless():

33 """Test that cuda_device_count_stateless changes return value if

34 - CUDA_VISIBLE_DEVICES is changed.

35 - """

34 + CUDA_VISIBLE_DEVICES is changed."""

35 if current_platform.is_rocm():

36 pytest.skip("Skip for ROCm because Ray uses HIP_VISIBLE_DEVICES.")

37 actor = _CUDADeviceCountStatelessTestActor.options( # type: ignore

tests/distributed/test_weight_transfer.py

+22 -44

@@ -1166,8 +1166,7 @@ class RecordingClient:

1166

1167 def _module_with(*pairs):

1168 """A tiny nn.Module exposing the given (name, tensor) pairs as parameters,

1169 - so trainer tests can build a ModuleSource without a real model.

1170 - """

1169 + so trainer tests can build a ModuleSource without a real model."""

1170 module = torch.nn.Module()

1171 for name, tensor in pairs:

1172 module.register_parameter(name, torch.nn.Parameter(tensor, requires_grad=False))

@@ -1288,8 +1287,7 @@ class TestModuleSource:

1287

1288 def test_metadata_agrees_with_iteration(self):

1289 """The two channels must line up element-for-element: engines declare

1291 - the round from `metadata()` and then send what iteration yields.

1292 - """

1290 + the round from `metadata()` and then send what iteration yields."""

1291 source = ModuleSource(

1292 _module_with(("w", torch.zeros(2, 3)), ("b", torch.zeros(3)))

1293 )

@@ -1303,13 +1301,11 @@ class TestModuleSource:

1301 class TestWeightSourceGroupContract:

1302 """`groups()` / `iter_groups()` on the WeightSource ABC. Groups are what

1303 backends gather and free by, so the default must agree with

1306 - `layerwise_groups` over `metadata()`, restricted to what this rank holds.

1307 - """

1304 + `layerwise_groups` over `metadata()`, restricted to what this rank holds."""

1305

1306 class _Source(WeightSource):

1307 """Minimal source over an ordered (name, tensor) list, optionally holding

1311 - only some names (in which case it iterates only their groups).

1312 - """

1308 + only some names (in which case it iterates only their groups)."""

1309

1310 def __init__(self, names, held=None, reverse=False):

1311 self._pairs = [(n, torch.full((2,), float(i))) for i, n in enumerate(names)]

@@ -1364,8 +1360,7 @@ class TestWeightSourceGroupContract:

1360 def test_groups_order_follows_metadata_not_the_declaration(self):

1361 """The declaration is a SET of names; the groups it selects still come

1362 out in metadata order, so each pairs with the right ``iter_groups()``

1367 - batch however the source listed them.

1368 - """

1363 + batch however the source listed them."""

1364 names = ["embed.w", "model.layers.0.a", "model.layers.1.a", "norm.w"]

1365 source = self._source(

1366 names, held=["model.layers.1.a", "model.layers.0.a", "model.layers.1.a"]

@@ -1399,16 +1394,14 @@ class TestWeightSourceGroupContract:

1394

1395 def test_out_of_order_iteration_raises(self):

1396 """Materializing is usually a collective, so a rank that iterates out of

1402 - order deadlocks its peers -- fail loudly instead.

1403 - """

1397 + order deadlocks its peers -- fail loudly instead."""

1398 source = self._source(["model.layers.0.a", "model.layers.0.b"], reverse=True)

1399 with pytest.raises(RuntimeError, match="iteration order must match"):

1400 list(source.iter_groups())

1401

1402 def test_a_source_may_override_iter_groups(self):

1403 """The extension point: materialize a whole group in one step instead of

1410 - one generator resume per tensor.

1411 - """

1404 + one generator resume per tensor."""

1405 calls = []

1406

1407 class _Batched(TestWeightSourceGroupContract._Source):

@@ -1429,8 +1422,7 @@ class TestDeferredProcessingContract:

1422 caller that takes over the update tail (running its own

1423 `finalize_layerwise_reload` instead of going through `finish_weight_update`)

1424 reads the flag and calls the method. Both must be answerable on any engine, or

1432 - that caller ends up reaching through a getattr.

1433 - """

1425 + that caller ends up reaching through a getattr."""

1426

1427 def _engines(self):

1428 return {

@@ -1444,8 +1436,7 @@ class TestDeferredProcessingContract:

1436

1437 def test_every_engine_can_be_drained(self):

1438 """The default is a no-op, so a caller never has to check whether the

1447 - method exists before calling it.

1448 - """

1439 + method exists before calling it."""

1440 for name, cls in self._engines().items():

1441 assert callable(cls.drain_pending), name

1442

@@ -1539,8 +1530,7 @@ class TestTrainerEngineBase:

1530 )

1531 def test_ipc_trainer_send_weights_drives_client_in_order():

1532 """send_weights issues start -> update -> finish and ships per-round metadata;

1542 - the packed wire param rides the init info, not the per-round update_info.

1543 - """

1533 + the packed wire param rides the init info, not the per-round update_info."""

1534 client = RecordingClient()

1535 engine = IPCTrainerWeightTransferEngine(

1536 client=client,

@@ -1559,8 +1549,7 @@ def test_ipc_trainer_send_weights_drives_client_in_order():

1549

1550 def test_ipc_trainer_init_ships_packed_to_worker():

1551 """trainer_init drives the inference-side init handshake and propagates the

1562 - must-agree `packed` flag to the worker.

1563 - """

1552 + must-agree `packed` flag to the worker."""

1553 if torch.accelerator.device_count() < 1:

1554 pytest.skip("Need at least 1 GPU (CUDA IPC handles).")

1555

@@ -1581,8 +1570,7 @@ def test_ipc_trainer_init_ships_packed_to_worker():

1570 def test_nccl_trainer_init_ships_worker_init_info(monkeypatch):

1571 """The sender's trainer_init drives the inference-side init handshake with

1572 the worker-shaped init info (rank_offset=1) while opening its own endpoint,

1584 - and propagates the must-agree wire params to the worker.

1585 - """

1573 + and propagates the must-agree wire params to the worker."""

1574 import vllm.distributed.weight_transfer.nccl_engine as nccl_engine_mod

1575

1576 # Bypass the real NCCL rendezvous.

@@ -1623,8 +1611,7 @@ def test_nccl_trainer_init_ships_worker_init_info(monkeypatch):

1611 def test_nccl_worker_learns_wire_params_from_init_handshake(monkeypatch):

1612 """The worker engine reads packed + buffer geometry from the

1613 trainer-supplied init info at the handshake, not from the config or the

1626 - per-round update info.

1627 - """

1614 + per-round update info."""

1615 import vllm.distributed.weight_transfer.nccl_engine as nccl_engine_mod

1616

1617 monkeypatch.setattr(

@@ -1657,8 +1644,7 @@ def test_nccl_worker_learns_wire_params_from_init_handshake(monkeypatch):

1644

1645 def test_nccl_trainer_init_non_sender_skips_rendezvous_and_client():

1646 """Non-sender trainer ranks build an engine without opening an endpoint or

1660 - touching the client; they only join the collectives in send_weights.

1661 - """

1647 + touching the client; they only join the collectives in send_weights."""

1648 client = RecordingClient()

1649 engine = WeightTransferTrainerFactory.trainer_init(

1650 init_info=NCCLTrainerInitInfo(

@@ -1688,8 +1674,7 @@ def test_nccl_trainer_init_non_sender_skips_rendezvous_and_client():

1674 def test_nccl_trainer_send_weights_drives_client_in_order():

1675 """send_weights issues start -> update -> finish and ships per-round

1676 metadata; the packed wire params ride the init handshake, not the

1691 - per-round update_info.

1692 - """

1677 + per-round update_info."""

1678 client = RecordingClient()

1679 engine = NCCLTrainerWeightTransferEngine(

1680 client=client,

@@ -1710,8 +1695,7 @@ def test_nccl_trainer_send_weights_drives_client_in_order():

1695

1696 class _ScriptedSource(WeightSource):

1697 """Declares `meta` but yields whatever `pairs` says — used to drive the

1713 - metadata/iteration agreement checks.

1714 - """

1698 + metadata/iteration agreement checks."""

1699

1700 def __init__(self, meta, pairs):

1701 self._meta = meta

@@ -1747,8 +1731,7 @@ def test_nccl_trainer_init_requires_source():

1731

1732 def test_nccl_trainer_send_weights_rejects_reordered_source(monkeypatch):

1733 """The worker sizes its buffers (and cuts packed chunks) from metadata(), so

1750 - iteration disagreeing with it must raise rather than corrupt the stream.

1751 - """

1734 + iteration disagreeing with it must raise rather than corrupt the stream."""

1735 meta = [

1736 ParamMeta("w", torch.float32, (4,)),

1737 ParamMeta("b", torch.float32, (2,)),

@@ -1762,8 +1745,7 @@ def test_nccl_trainer_send_weights_rejects_reordered_source(monkeypatch):

1745

1746 def test_nccl_trainer_send_weights_rejects_dtype_disagreement(monkeypatch):

1747 """A source that declares one wire dtype and materializes another would make

1765 - the two sides disagree on every byte offset.

1766 - """

1748 + the two sides disagree on every byte offset."""

1749 meta = [ParamMeta("w", torch.float32, (4,))]

1750 engine = _mock_group_engine(

1751 _ScriptedSource(meta, [("w", torch.zeros(4, dtype=torch.bfloat16))]),

@@ -1790,8 +1772,7 @@ def test_nccl_trainer_send_weights_rejects_truncated_source(monkeypatch):

1772

1773 def test_nccl_trainer_send_weights_broadcasts_contiguous(monkeypatch):

1774 """NCCL sends numel elements from data_ptr(), so a non-contiguous view must

1793 - be linearized first or the worker receives unrelated memory.

1794 - """

1775 + be linearized first or the worker receives unrelated memory."""

1776 base = torch.arange(6, dtype=torch.float32).reshape(2, 3)

1777 view = base.t() # non-contiguous

1778 meta = [ParamMeta("w", torch.float32, tuple(view.shape))]

@@ -1861,8 +1842,7 @@ def _sparse_patch(device: str = "cpu") -> SparseWeightPatch:

1842 def test_sparse_nccl_trainer_init_ships_worker_init_info(monkeypatch):

1843 """The sender's trainer_init drives the init handshake with the

1844 worker-shaped init info; sparse ships no packed wire params, so the worker

1864 - keeps its unpacked defaults. Sparse takes no `source`.

1865 - """

1845 + keeps its unpacked defaults. Sparse takes no `source`."""

1846 import vllm.distributed.weight_transfer.sparse_nccl_engine as sparse_mod

1847

1848 monkeypatch.setattr(sparse_mod, "open_trainer_endpoint", lambda info: MagicMock())

@@ -1915,8 +1895,7 @@ def test_sparse_nccl_trainer_send_weights_drives_client_in_order(monkeypatch):

1895

1896 def test_sparse_nccl_trainer_send_weights_empty_round_is_noop():

1897 """A round with no patches must not touch the client (an empty sparse

1918 - update info is invalid by construction).

1919 - """

1898 + update info is invalid by construction)."""

1899 client = RecordingClient()

1900 engine = SparseNCCLTrainerWeightTransferEngine(client=client)

1901 engine.model_update_group = MagicMock()

@@ -1952,8 +1931,7 @@ def test_sparse_nccl_trainer_rejects_source():

1931 def test_sparse_nccl_trainer_validates_patch_before_any_rpc():

1932 """Malformed patches must fail on the trainer, before start_weight_update:

1933 the worker's own checks only run once the broadcasts are already in flight,

1955 - where a size mismatch wedges both sides instead of raising.

1956 - """

1934 + where a size mismatch wedges both sides instead of raising."""

1935 client = RecordingClient()

1936 engine = SparseNCCLTrainerWeightTransferEngine(client=client)

1937 engine.model_update_group = MagicMock()

tests/distributed/test_weight_transfer_nccl_uid.py

+3 -6

@@ -403,8 +403,7 @@ def trainer_broadcast_tensor_torch_free(

403 def _run_uid_transfer(trainer_task) -> None:

404 """Mint a unique id, run the worker and the given trainer peer concurrently,

405 and assert the tensor round-trips. Both ranks must enter init together --

406 - there is no store barrier on the UID path.

407 - """

406 + there is no store barrier on the UID path."""

407 from vllm.distributed.device_communicators.pynccl_wrapper import NCCLLibrary

408

409 _init_ray_for_weight_transfer()

@@ -451,8 +450,7 @@ def _run_uid_transfer(trainer_task) -> None:

450 )

451 def test_nccl_weight_transfer_between_processes_uid():

452 """Weight transfer over a pre-shared ncclUniqueId (no TCPStore), with both

454 - ranks on vLLM's ``PyNcclCommunicator``.

455 - """

453 + ranks on vLLM's ``PyNcclCommunicator``."""

454 _run_uid_transfer(trainer_broadcast_tensor_uid)

455

456

@@ -462,7 +460,6 @@ def test_nccl_weight_transfer_between_processes_uid():

460 )

461 def test_nccl_weight_transfer_torch_free_trainer():

462 """UID weight transfer against a foreign peer that uses no torch and no vLLM

465 - communicator, only a cupy buffer and raw ``ncclBroadcast``.

466 - """

463 + communicator, only a cupy buffer and raw ``ncclBroadcast``."""

464 pytest.importorskip("cupy")

465 _run_uid_transfer(trainer_broadcast_tensor_torch_free)

tests/engine/test_arg_utils.py

+5 -10

@@ -570,8 +570,7 @@ def test_attention_config():

570 def test_multi_node_world_size_includes_pcp(monkeypatch):

571 """PCP expands the process world size, so the --nnodes divisibility check

572 must include it. Without this, TP=1/PCP=2 over 2 nodes computes a world

573 - size of 1 and the launch is rejected before the engine starts.

574 - """

573 + size of 1 and the launch is rejected before the engine starts."""

574 import vllm.config.vllm

575

576 # PCP requires the V2 model runner, which is gated on Triton.

@@ -797,8 +796,7 @@ def test_expand_json_human_readable_numbers(input_json, expected_json):

796 )

797 def test_cloud_storage_uri_skips_get_model_path(uri, monkeypatch):

798 """Cloud storage URIs should not be passed to get_model_path()

800 - when HF_HUB_OFFLINE=1, as they are not valid HF repo IDs.

801 - """

799 + when HF_HUB_OFFLINE=1, as they are not valid HF repo IDs."""

800 import huggingface_hub

801

802 monkeypatch.setattr(huggingface_hub.constants, "HF_HUB_OFFLINE", True)

@@ -810,8 +808,7 @@ def test_cloud_storage_uri_skips_get_model_path(uri, monkeypatch):

808

809 def test_cloud_storage_tokenizer_skips_get_model_path(monkeypatch):

810 """Cloud storage tokenizer URI should not be passed to

813 - get_model_path() when HF_HUB_OFFLINE=1.

814 - """

811 + get_model_path() when HF_HUB_OFFLINE=1."""

812 import huggingface_hub

813

814 monkeypatch.setattr(huggingface_hub.constants, "HF_HUB_OFFLINE", True)

@@ -973,8 +970,7 @@ class TestDpDeviceIdSharding:

970 def test_dp_supervisor_device_ids_stay_env_relative(self):

971 """Regression test: the DP supervisor must pass env-relative indices,

972 not physical IDs, because each child re-resolves --device-ids

976 - against its inherited device-control env var.

977 - """

973 + against its inherited device-control env var."""

974 import argparse

975

976 from vllm.entrypoints.launchers.dp_supervisor import _build_device_ids

@@ -1001,8 +997,7 @@ class TestDpDeviceIdSharding:

997

998 def test_dp_rank_shards_user_assigned_gpu_ids(self):

999 """get_physical_gpu_ids_for_local_dp_rank slices the user-provided

1004 - --device-ids list instead of recomputing from the env var.

1005 - """

1000 + --device-ids list instead of recomputing from the env var."""

1001 from vllm.platforms import current_platform

1002 from vllm.v1.engine.utils import get_physical_gpu_ids_for_local_dp_rank

1003

tests/entrypoints/anthropic/test_anthropic_messages_conversion.py

+10 -20

@@ -241,8 +241,7 @@ class TestToolResultContent:

241 self, tool_result_content

242 ) -> AnthropicMessagesRequest:

243 """Build a request with assistant tool_use followed by user

244 - tool_result.

245 - """

244 + tool_result."""

245 return _make_request(

246 [

247 {

@@ -327,8 +326,7 @@ class TestToolResultContent:

326

327 def test_tool_result_with_text_and_image(self):

328 """Mixed text+image tool_result: text in tool msg, image in user

330 - msg.

331 - """

329 + msg."""

330 request = self._make_tool_result_request(

331 [

332 {"type": "text", "text": "Here is the screenshot"},

@@ -425,8 +423,7 @@ class TestToolResultContent:

423 class TestAttributionHeaderStripping:

424 def test_billing_header_stripped_from_system(self):

425 """Claude Code's x-anthropic-billing-header block should be

428 - stripped to preserve prefix caching.

429 - """

426 + stripped to preserve prefix caching."""

427 request = _make_request(

428 [{"role": "user", "content": "Hello"}],

429 system=[

@@ -1150,8 +1147,7 @@ class TestMessageStreamConverterToolUseContentBuffering:

1147 async def test_buffered_content_flushed_on_done_without_usage_chunk(self):

1148 """Content buffered during tool_use must be emitted even if the

1149 stream jumps straight from finish_reason to [DONE], skipping the

1153 - empty-choices usage chunk.

1154 - """

1150 + empty-choices usage chunk."""

1151

1152 async def sse_input():

1153 yield _make_stream_chunk(

@@ -1255,8 +1251,7 @@ class TestStreamingCacheUsageSemantics:

1251 async def test_streaming_cache_fields_absent_then_populated(self):

1252 """First chunk lacks prompt_tokens_details (vLLM contract);

1253 message_start omits cache fields. The final chunk carries

1258 - prompt_tokens_details, so message_delta carries resolved values.

1259 - """

1254 + prompt_tokens_details, so message_delta carries resolved values."""

1255

1256 async def sse_input():

1257 yield _make_stream_chunk(

@@ -1301,8 +1296,7 @@ class TestStreamingCacheUsageSemantics:

1296 @pytest.mark.asyncio

1297 async def test_streaming_no_cache_hit(self):

1298 """When the final chunk reports cached_tokens=0, message_delta carries

1304 - cache fields = 0 (cache miss); message_start still omits them.

1305 - """

1299 + cache fields = 0 (cache miss); message_start still omits them."""

1300

1301 async def sse_input():

1302 yield _make_stream_chunk(

@@ -1343,8 +1337,7 @@ class TestStreamingCacheUsageSemantics:

1337 @pytest.mark.asyncio

1338 async def test_streaming_no_prompt_tokens_details_at_all(self):

1339 """If --enable-prompt-tokens-details is off, no chunk carries cache

1346 - info; both message_start and message_delta omit cache fields.

1347 - """

1340 + info; both message_start and message_delta omit cache fields."""

1341

1342 async def sse_input():

1343 yield _make_stream_chunk(

@@ -1465,8 +1458,7 @@ class TestMessagesFullConverter:

1458 class TestCacheSalt:

1459 def test_cache_salt_passed_through(self):

1460 """cache_salt on the Anthropic request reaches the converted

1468 - ChatCompletionRequest so prefix-cache isolation works via /v1/messages.

1469 - """

1461 + ChatCompletionRequest so prefix-cache isolation works via /v1/messages."""

1462 request = _make_request(

1463 [{"role": "user", "content": "Hello"}],

1464 cache_salt="tenant-abc-secret-salt",

@@ -1671,8 +1663,7 @@ class TestClientErrorResponses:

1663 @staticmethod

1664 def _conversion_error() -> ValidationError:

1665 """A real pydantic ValidationError like the one ChatCompletionRequest

1674 - construction raises when Anthropic input violates the OpenAI schema.

1675 - """

1666 + construction raises when Anthropic input violates the OpenAI schema."""

1667

1668 class _StubRequest(BaseModel):

1669 stop: Annotated[list[str], Field(max_length=4)] | None = None

@@ -1683,8 +1674,7 @@ class TestClientErrorResponses:

1674

1675 def test_validation_error_returns_bad_request(self):

1676 """A pydantic ValidationError during Anthropic->OpenAI conversion is

1686 - surfaced as a 400 BadRequestError, not a 500.

1687 - """

1677 + surfaced as a 400 BadRequestError, not a 500."""

1678 handler = MagicMock(spec=AnthropicServingMessages)

1679 handler.create_messages.side_effect = self._conversion_error()

1680

tests/entrypoints/cohere/test_serving_streaming.py

+2 -4

@@ -87,8 +87,7 @@ def _make_chunk(

87 citations: list[Any] | None = None,

88 ) -> str:

89 """Build the ``data: {...}\\n\\n`` SSE frame the production code

90 - consumes.

91 - """

90 + consumes."""

91 delta: dict[str, Any] = {}

92 if role is not None:

93 delta["role"] = role

@@ -136,8 +135,7 @@ async def _stream_from(items: list[str]) -> AsyncGenerator[str, None]:

135

136 async def _drain(serving: CohereServingChatV2, items: list[str]) -> list[str]:

137 """Drive ``_chat_completion_stream_to_v2`` over ``items`` and collect

139 - the emitted SSE frames.

140 - """

138 + the emitted SSE frames."""

139 request = CohereChatV2Request(

140 model="m", messages=[{"role": "user", "content": "hi"}], stream=True

141 )

tests/entrypoints/launchers/api_server/test_api_server_process_manager.py

+4 -8

@@ -50,8 +50,7 @@ def update_admission_stats_worker(listen_address, sock, args, client_config):

50 # importable by `multiprocessing.spawn` (no closures, no nesting).

51 def defer_addresses_stub_worker(listen_address, sock, args, client_config):

52 """Bind ROUTER/PULL with a kernel-assigned port, report the actual

53 - endpoints back via the pipe, then exit.

54 - """

53 + endpoints back via the pipe, then exit."""

54 ctx = zmq.Context()

55 try:

56 in_sock = make_zmq_socket(

@@ -358,8 +357,7 @@ def test_external_process_monitoring(api_server_args):

357 def test_gather_actual_addresses_end_to_end():

358 """Each child binds ROUTER/PULL with a kernel-picked port and reports

359 the bound endpoints back via its per-child pipe; the manager surfaces

361 - them via :py:meth:`gather_actual_addresses`.

362 - """

360 + them via :py:meth:`gather_actual_addresses`."""

361 host = "127.0.0.1"

362 num_servers = 4

363

@@ -409,8 +407,7 @@ def test_gather_actual_addresses_end_to_end():

407 @pytest.mark.timeout(30)

408 def test_gather_actual_addresses_child_crash_before_report():

409 """A child that exits before sending its endpoints must surface a

412 - clear ``RuntimeError`` rather than hang or return ``None`` slots.

413 - """

410 + clear ``RuntimeError`` rather than hang or return ``None`` slots."""

411 host = "127.0.0.1"

412 num_servers = 2

413 placeholder_inputs = [

@@ -447,8 +444,7 @@ def test_gather_actual_addresses_child_crash_before_report():

444

445 def test_rust_frontend_launch_log_redacts_credentials(monkeypatch, caplog):

446 """The Rust frontend command carries every non-default arg as JSON.

450 - Credentials must not reach the log line.

451 - """

447 + Credentials must not reach the log line."""

448 import subprocess as subprocess_mod

449

450 from vllm.entrypoints.launchers.cli_args import make_arg_parser

tests/entrypoints/launchers/test_cli_args.py

+2 -4

@@ -186,8 +186,7 @@ def test_enable_auto_choice_fails_with_enable_reasoning(serve_parser):

186

187 def test_passes_with_reasoning_parser(serve_parser):

188 """Ensure validation passes if reasoning is enabled

189 - with a reasoning parser

190 - """

189 + with a reasoning parser"""

190 args = serve_parser.parse_args(

191 args=[

192 "--reasoning-parser",

@@ -382,8 +381,7 @@ def test_default_chat_template_kwargs_invalid_json(serve_parser):

381 )

382 def test_served_model_name_parsing(tmp_path, vllm_parser, args, raises):

383 """Ensure that users don't misuse --served-model-name and end up with the default

385 - model tag instead of the one they intended to serve.

386 - """

384 + model tag instead of the one they intended to serve."""

385 # Call the serve subparser

386 args.insert(0, "serve")

387 # Create a dummy config file if the test case includes it

tests/entrypoints/multimodal/openai/chat_completion/test_chat_completion_with_mixed_audio_embeds.py

+1 -2

@@ -2,8 +2,7 @@

2 # SPDX-FileCopyrightText: Copyright contributors to the vLLM project

3

4 """E2E test for mixing `prompt_embeds` with `audio_embeds` in a single

5 -Chat Completions request.

6 -"""

5 +Chat Completions request."""

6

7 import json

8

tests/entrypoints/multimodal/openai/chat_completion/test_chat_completion_with_mixed_image_embeds.py

+1 -2

@@ -68,8 +68,7 @@ async def client(server):

68 def image_url() -> str:

69 """Stable real image as a data URL, kept identical across both the

70 text and prompt_embeds requests so any output difference must come from

71 - how the text content is delivered.

72 - """

71 + how the text content is delivered."""

72 return encode_image_url(ImageAsset("stop_sign").pil_image)

73

74

tests/entrypoints/multimodal/openai/chat_completion/test_vision.py

+2 -4

@@ -145,8 +145,7 @@ def describe_image_messages(

145 ) -> list[dict]:

146 """Build the system + user messages used by the completions-with-image

147 family of tests. *extra_image_fields* is merged into the top-level

148 - image content block (for uuid / bad-key tests).

149 - """

148 + image content block (for uuid / bad-key tests)."""

149 image_block: dict = {

150 "type": "image_url",

151 "image_url": {"url": image_url},

@@ -176,8 +175,7 @@ async def complete_and_check(

175 temperature: float = 0.0,

176 ) -> str:

177 """Run a chat completion and assert the output is non-empty.

179 - Returns the content string.

180 - """

178 + Returns the content string."""

179 chat_completion = await client.chat.completions.create(

180 model=model_name,

181 messages=messages,

tests/entrypoints/openai/chat_completion/test_chat_error.py

+2 -4

@@ -465,8 +465,7 @@ def test_system_message_warns_on_video(video_content):

465 def test_json_schema_response_format_missing_schema():

466 """When response_format type is 'json_schema' but the json_schema field

467 is not provided, request construction should raise a validation error

468 - so the API returns 400 instead of 500.

469 - """

468 + so the API returns 400 instead of 500."""

469 with pytest.raises(Exception, match="json_schema.*must be provided"):

470 ChatCompletionRequest(

471 model=MODEL_NAME,

@@ -520,8 +519,7 @@ def test_structured_outputs_structural_tag_invalid(structural_tag):

519 @pytest.mark.parametrize("field_name", ["prompt_logprobs", "top_logprobs"])

520 def test_non_numeric_logprobs_rejected(field_name):

521 """A non-numeric logprobs value must be a clean 400 validation error, not a

523 - TypeError from the mode='before' comparison (which surfaces as HTTP 500).

524 - """

522 + TypeError from the mode='before' comparison (which surfaces as HTTP 500)."""

523 with pytest.raises(VLLMValidationError, match=f"`{field_name}` must be an integer"):

524 ChatCompletionRequest(

525 model=MODEL_NAME,

tests/entrypoints/openai/chat_completion/test_chat_logit_bias_validation.py

+3 -6

@@ -81,8 +81,7 @@ async def test_chat_logit_bias_invalid(client):

81 @pytest.mark.asyncio

82 async def test_chat_logit_bias_non_integer_key(client):

83 """Test that a non-integer logit_bias key is rejected with a clean,

84 - informative error instead of a raw 'invalid literal for int()' message.

85 - """

84 + informative error instead of a raw 'invalid literal for int()' message."""

85 with pytest.raises(openai.BadRequestError) as excinfo:

86 await client.chat.completions.create(

87 model=MODEL_NAME,

@@ -102,8 +101,7 @@ async def test_chat_logit_bias_non_integer_key(client):

101 @pytest.mark.asyncio

102 async def test_chat_logit_bias_non_numeric_value(client):

103 """Test that a non-numeric logit_bias value is rejected with a message

105 - that names the specific offending token, not just a generic TypeError.

106 - """

104 + that names the specific offending token, not just a generic TypeError."""

105 with pytest.raises(openai.BadRequestError) as excinfo:

106 await client.chat.completions.create(

107 model=MODEL_NAME,

@@ -122,8 +120,7 @@ async def test_chat_logit_bias_non_numeric_value(client):

120 @pytest.mark.asyncio

121 async def test_chat_logit_bias_multiple_non_integer_keys(client):

122 """Test that ALL invalid logit_bias keys are reported together,

125 - not just the first one encountered.

126 - """

123 + not just the first one encountered."""

124 with pytest.raises(openai.BadRequestError) as excinfo:

125 await client.chat.completions.create(

126 model=MODEL_NAME,

tests/entrypoints/openai/chat_completion/test_extra_content_fields.py

+2 -4

@@ -54,8 +54,7 @@ def stop_sign_image_url():

54 async def test_translategemma_extra_lang_code_fields(client: openai.AsyncOpenAI):

55 """En -> es translation through TranslateGemma's bundled chat template's

56 ``text`` content branch, which depends on ``source_lang_code`` /

57 - ``target_lang_code`` extra fields being preserved on the content part.

58 - """

57 + ``target_lang_code`` extra fields being preserved on the content part."""

58 completion = await client.chat.completions.create(

59 model=MODEL_NAME,

60 messages=[

@@ -99,8 +98,7 @@ async def test_translategemma_image_extra_lang_code_fields(

98 ``source_lang_code`` / ``target_lang_code`` extras must survive parsing

99 and end up alongside the ``{"type": "image"}`` placeholder so the

100 template's image branch (``content["type"] == 'image'``) renders the

102 - right OCR-translation prompt.

103 - """

101 + right OCR-translation prompt."""

102 completion = await client.chat.completions.create(

103 model=MODEL_NAME,

104 messages=[

tests/entrypoints/openai/chat_completion/test_logprob_token_ids.py

+1 -2

@@ -191,8 +191,7 @@ async def test_logprob_token_ids_stream_with_no_top_k(server):

191 async def test_logprob_token_ids_default_behavior_unchanged(server):

192 """Without `logprob_token_ids`, the response carries the natural top-k

193 most-likely tokens. This guards against the new field accidentally

194 - changing the default-path output.

195 - """

194 + changing the default-path output."""

195 async with server.get_async_client() as client:

196 resp = await client.chat.completions.create(

197 model=MODEL_NAME,

tests/entrypoints/openai/chat_completion/test_serving_chat.py

+4 -8

@@ -1406,8 +1406,7 @@ async def test_serving_chat_did_set_correct_cache_salt(model_type):

1406 @pytest.mark.asyncio

1407 async def test_serving_chat_data_parallel_rank_extraction():

1408 """Test that data_parallel_rank is properly extracted from header and

1409 - passed to engine.

1410 - """

1409 + passed to engine."""

1410 mock_engine = MagicMock(spec=AsyncLLM)

1411 mock_engine.errored = False

1412 mock_engine.model_config = MockModelConfig()

@@ -1729,8 +1728,7 @@ class TestServingChatWithHarmony:

1728 @pytest.mark.asyncio

1729 async def test_system_message_without_tools(self, serving_chat, stream):

1730 """Leading system message produces a developer message with

1732 - DeveloperContent (# Instructions header).

1733 - """

1731 + DeveloperContent (# Instructions header)."""

1732 messages = [

1733 {"role": "system", "content": "You are a helpful assistant."},

1734 {"role": "user", "content": "Hello"},

@@ -1752,8 +1750,7 @@ class TestServingChatWithHarmony:

1750 @pytest.mark.asyncio

1751 async def test_system_message_with_tools(self, serving_chat, stream, weather_tools):

1752 """Leading system message is folded into the developer message

1755 - alongside tool definitions.

1756 - """

1753 + alongside tool definitions."""

1754 messages = [

1755 {"role": "system", "content": "You are a helpful assistant."},

1756 {"role": "user", "content": "What's the weather?"},

@@ -2168,8 +2165,7 @@ class TestServingChatWithHarmony:

2165 @pytest.mark.asyncio

2166 async def test_tool_choice_validation_without_parser():

2167 """Test that tool_choice='required' or named tool without tool_parser

2171 - returns an appropriate error message.

2172 - """

2168 + returns an appropriate error message."""

2169 mock_engine = MagicMock(spec=AsyncLLM)

2170 mock_engine.errored = False

2171 mock_engine.model_config = MockModelConfig()

tests/entrypoints/openai/chat_completion/test_thinking_token_budget.py

+1 -2

@@ -151,8 +151,7 @@ async def client(request, server, server_with_auto_reasoning_config):

151 @pytest.mark.parametrize("client", ["default", "auto_config"], indirect=True)

152 async def test_thinking_token_budget_mixed_requests(client: openai.AsyncOpenAI):

153 """Test that mixed requests (some with thinking_token_budget, some without)

154 - complete successfully without errors.

155 - """

154 + complete successfully without errors."""

155 response_with_budget = await client.chat.completions.create(

156 model=MODEL_NAME,

157 messages=MESSAGES,

tests/entrypoints/openai/completion/test_completion_error.py

+3 -6

@@ -521,8 +521,7 @@ async def test_completion_error_stream():

521 def test_json_schema_response_format_missing_schema():

522 """When response_format type is 'json_schema' but the json_schema field

523 is not provided, request construction should raise a validation error

524 - so the API returns 400 instead of 500.

525 - """

524 + so the API returns 400 instead of 500."""

525 with pytest.raises(Exception, match="json_schema.*must be provided"):

526 CompletionRequest(

527 model=MODEL_NAME,

@@ -585,8 +584,7 @@ def test_negative_prompt_token_ids_flat():

584 def test_logprobs_minus_one_allowed():

585 """logprobs=-1 means "return all logprobs". The sampling layer and the chat

586 top_logprobs / prompt_logprobs validators all accept -1, so the completion

588 - logprobs validator must accept it too instead of rejecting it as negative.

589 - """

587 + logprobs validator must accept it too instead of rejecting it as negative."""

588 request = CompletionRequest(

589 model=MODEL_NAME,

590 prompt="Test prompt",

@@ -746,8 +744,7 @@ class TestCompletionPromptListLimit:

744 @pytest.mark.parametrize("field_name", ["prompt_logprobs", "logprobs"])

745 def test_non_numeric_logprobs_rejected(field_name):

746 """A non-numeric logprobs value must be a clean 400 validation error, not a

749 - TypeError from the mode='before' comparison (which surfaces as HTTP 500).

750 - """

747 + TypeError from the mode='before' comparison (which surfaces as HTTP 500)."""

748 with pytest.raises(VLLMValidationError, match=f"`{field_name}` must be an integer"):

749 CompletionRequest(

750 model=MODEL_NAME,

tests/entrypoints/openai/parser/test_harmony_render_parity.py

+4 -8

@@ -43,8 +43,7 @@ def _system():

43

44 class TestResponseInputToHarmonyRenderParity:

45 """Each test drives the same conversation through both APIs and asserts

46 - identical Harmony messages and rendered token sequences.

47 - """

46 + identical Harmony messages and rendered token sequences."""

47

48 # -----------------------------------------------------------------------

49 # Single-message cases

@@ -52,8 +51,7 @@ class TestResponseInputToHarmonyRenderParity:

51

52 def test_developer_message(self):

53 """Both APIs must render developer messages identically using

55 - DeveloperContent (with the '# Instructions' header).

56 - """

54 + DeveloperContent (with the '# Instructions' header)."""

55 chat_msgs = parse_chat_input_to_harmony_message(

56 {"role": "developer", "content": "Be concise."}

57 )

@@ -248,8 +246,7 @@ class TestResponseInputToHarmonyRenderParity:

246 def test_reasoning_combined_with_function_call(self):

247 """Chat API packs reasoning + tool_calls into one dict; responses API

248 represents them as two separate items. Both must produce the same two

251 - Harmony messages in the same order: analysis then commentary.

252 - """

249 + Harmony messages in the same order: analysis then commentary."""

250 chat_msgs = parse_chat_input_to_harmony_message(

251 {

252 "role": "assistant",

@@ -532,8 +529,7 @@ class TestResponseInputToHarmonyRenderParity:

529 def test_completed_turns_drop_reasoning(self):

530 """Validates that reasoning from completed turns is dropped, while

531 reasoning from the current in-progress tool-call turn is preserved

535 - in both chat completions and responses previous_input_messages.

536 - """

532 + in both chat completions and responses previous_input_messages."""

533 first_turn_reasoning = "FIRST_TURN_REASONING"

534 second_turn_reasoning = "SECOND_TURN_REASONING"

535

tests/entrypoints/openai/parser/test_harmony_utils.py

+1 -2

@@ -323,8 +323,7 @@ class TestCommonParseInputToHarmonyMessage:

323

324 def test_system_message(self, parse_function):

325 """Test parsing system messages, which are parsed into developer messages

326 - with DeveloperContent.

327 - """

326 + with DeveloperContent."""

327 chat_msg = {

328 "role": "system",

329 "content": "You are a helpful assistant",

tests/entrypoints/openai/responses/test_function_call_parsing.py

+3 -6

@@ -12,8 +12,7 @@ from vllm.entrypoints.openai.responses.protocol import ResponsesRequest

12

13 def test_function_call_dict_converted_to_object():

14 """Test that function_call dictionaries are correctly parsed into

15 - ResponseFunctionToolCall objects.

16 - """

15 + ResponseFunctionToolCall objects."""

16 # Create a request with function_call as dict

17 request_data = {

18 "model": "gpt-oss",

@@ -330,8 +329,7 @@ def test_validator_handles_empty_iterator():

329

330 def test_assistant_string_content_stays_easyinput():

331 """EasyInput assistant message with plain string content is not

333 - coerced into a ResponseOutputMessage.

334 - """

332 + coerced into a ResponseOutputMessage."""

333 request_data = {

334 "model": "test-model",

335 "input": [

@@ -353,8 +351,7 @@ def test_assistant_string_content_stays_easyinput():

351

352 def test_assistant_output_style_content_coerced():

353 """Assistant message whose content is output-message-shaped (list of

356 - output_text items) should be coerced to ResponseOutputMessage.

357 - """

354 + output_text items) should be coerced to ResponseOutputMessage."""

355 request_data = {

356 "model": "test-model",

357 "input": [

tests/entrypoints/openai/responses/test_response_input_to_harmony.py

+4 -8

@@ -67,8 +67,7 @@ class TestResponseInputToHarmonyMessage:

67

68 def test_system_message(self):

69 """System messages carry developer instructions and must be rendered

70 - as developer messages with DeveloperContent.

71 - """

70 + as developer messages with DeveloperContent."""

71 msg = response_input_to_harmony(

72 {"type": "message", "role": "system", "content": "Be helpful."},

73 prev_responses=[],

@@ -90,8 +89,7 @@ class TestResponseInputToHarmonyMessage:

89

90 def test_developer_message_gets_instructions_prefix(self):

91 """Developer messages must use DeveloperContent which adds the

93 - '# Instructions' header the model was trained on.

94 - """

92 + '# Instructions' header the model was trained on."""

93 msg = response_input_to_harmony(

94 {"type": "message", "role": "developer", "content": "Be concise."},

95 prev_responses=[],

@@ -121,8 +119,7 @@ class TestResponseInputToHarmonyMessage:

119

120 def test_developer_message_array_content_concatenated(self):

121 """Array content in developer messages is flattened and rendered

124 - via DeveloperContent with the '# Instructions' header.

125 - """

122 + via DeveloperContent with the '# Instructions' header."""

123 msg = response_input_to_harmony(

124 {

125 "type": "message",

@@ -213,8 +210,7 @@ class TestResponseInputToHarmonyMessage:

210

211 def test_function_call_output_uses_most_recent_matching_call(self):

212 """When multiple prev_responses share a call_id, the last one wins

216 - because the search is reversed.

217 - """

213 + because the search is reversed."""

214 earlier = ResponseFunctionToolCall(

215 id="fc_old",

216 call_id="call_test",

tests/entrypoints/openai/responses/test_responses_utils.py

+3 -6

@@ -839,13 +839,11 @@ class TestConstructChatMessagesCombinePolicy:

839

840 class TestConstructInputMessagesInstructionsLeak:

841 """Regression tests for #37697: instructions from a prior response

842 - should NOT leak through previous_response_id.

843 - """

842 + should NOT leak through previous_response_id."""

843

844 def test_old_instructions_stripped_from_prev_msg(self):

845 """System message in prev_msg must be dropped so the new request's

847 - instructions are the only system message in the conversation.

848 - """

846 + instructions are the only system message in the conversation."""

847 prev = [

848 {"role": "system", "content": "old instructions"},

849 {"role": "user", "content": "What is 2+2?"},

@@ -862,8 +860,7 @@ class TestConstructInputMessagesInstructionsLeak:

860

861 def test_no_instructions_in_new_request(self):

862 """If the new request has no instructions, old ones should still

865 - be stripped -- they must not carry over.

866 - """

863 + be stripped -- they must not carry over."""

864 prev = [

865 {"role": "system", "content": "old instructions"},

866 {"role": "user", "content": "Hi"},

tests/entrypoints/openai/responses/test_serving_responses.py

+6 -12

@@ -1163,8 +1163,7 @@ class TestHarmonyPreambleStreaming:

1163

1164 def test_preamble_done_emits_text_done_events(self) -> None:

1165 """Completed preamble should emit text done + content_part done +

1166 - output_item done, same shape as final channel.

1167 - """

1166 + output_item done, same shape as final channel."""

1167 from vllm.entrypoints.openai.responses.streaming_events import (

1168 emit_previous_item_done_events,

1169 )

@@ -1185,8 +1184,7 @@ class TestHarmonyPreambleStreaming:

1184

1185 def test_commentary_with_recipient_no_preamble_done(self) -> None:

1186 """Commentary + recipient='functions.X' should route to function call

1188 - done, not preamble done.

1189 - """

1187 + done, not preamble done."""

1188 from vllm.entrypoints.openai.responses.streaming_events import (

1189 emit_previous_item_done_events,

1190 )

@@ -1332,8 +1330,7 @@ def _mock_parser_with_reasoning(serving, delta_sequence: list[DeltaMessage]):

1330 class TestStreamingReasoningToContentTransition:

1331 """Tests for _process_simple_streaming_events reasoning-to-content

1332 transition, specifically the fix for mixed deltas that carry both

1335 - reasoning and content simultaneously.

1336 - """

1333 + reasoning and content simultaneously."""

1334

1335 @pytest.mark.asyncio

1336 async def test_mixed_delta_reasoning_and_content_emits_reasoning_delta(

@@ -1343,8 +1340,7 @@ class TestStreamingReasoningToContentTransition:

1340 and content set (e.g. reasoning end and content start in the same

1341 chunk), the trailing reasoning text must be emitted as a

1342 ResponseReasoningTextDeltaEvent and included in the

1346 - ResponseReasoningTextDoneEvent text.

1347 - """

1343 + ResponseReasoningTextDoneEvent text."""

1344 monkeypatch.setattr(envs, "VLLM_USE_EXPERIMENTAL_PARSER_CONTEXT", False)

1345 serving = _make_serving_instance_with_reasoning()

1346

@@ -1413,8 +1409,7 @@ class TestStreamingReasoningToContentTransition:

1409 self, monkeypatch

1410 ):

1411 """When the transition from reasoning to content is clean (no mixed

1416 - delta), no extra reasoning delta event should be emitted.

1417 - """

1412 + delta), no extra reasoning delta event should be emitted."""

1413 monkeypatch.setattr(envs, "VLLM_USE_EXPERIMENTAL_PARSER_CONTEXT", False)

1414 serving = _make_serving_instance_with_reasoning()

1415

@@ -1475,8 +1470,7 @@ class TestStreamingReasoningToContentTransition:

1470 async def test_reasoning_only_stream_no_content(self, monkeypatch):

1471 """When the stream has only reasoning deltas and no content, the

1472 reasoning done event should be emitted at finalization with the

1478 - full accumulated text, and no text delta events should appear.

1479 - """

1473 + full accumulated text, and no text delta events should appear."""

1474 monkeypatch.setattr(envs, "VLLM_USE_EXPERIMENTAL_PARSER_CONTEXT", False)

1475 serving = _make_serving_instance_with_reasoning()

1476

tests/entrypoints/openai/responses/test_simple.py

+1 -2

@@ -143,8 +143,7 @@ async def test_streaming_output_consistency(client: OpenAI, model_name: str):

143 @pytest.mark.parametrize("model_name", [MODEL_NAME])

144 async def test_streaming_logprobs(client: OpenAI, model_name: str):

145 """Test that streaming with logprobs returns valid logprob data on

146 - output_text.delta events and that top_logprobs has the requested count.

147 - """

146 + output_text.delta events and that top_logprobs has the requested count."""

147 response = await client.responses.create(

148 model=model_name,

149 input="Say hello.",

tests/entrypoints/openai/responses/test_streaming_events.py

+2 -4

@@ -124,8 +124,7 @@ class TestProcessorCompoundDeltas:

124

125 def test_split_name_and_args_same_index(self):

126 """Regression: parsers like KimiK2 emit name and args as separate

127 - DeltaToolCalls at the same index within one DeltaMessage.

128 - """

127 + DeltaToolCalls at the same index within one DeltaMessage."""

128 tc_name = _make_tool_call(0, name="get_weather")

129 tc_args = _make_tool_call(0, arguments='{"city":"SF"}')

130 delta = DeltaMessage(tool_calls=[tc_name, tc_args])

@@ -141,8 +140,7 @@ class TestProcessorCompoundDeltas:

140

141 def test_reasoning_to_content_transition(self):

142 """Regression: the old special case in emit_delta handled this;

144 - now split_delta handles it generically.

145 - """

143 + now split_delta handles it generically."""

144 processor = SimpleStreamingEventProcessor()

145 _run_through_processor(processor, DeltaMessage(reasoning="think"))

146 assert processor.state.current_state == _StateType.REASONING

tests/entrypoints/openai/test_render_token_offsets.py

+3 -6

@@ -25,8 +25,7 @@ def _model_config() -> Mock:

25 def test_completion_flag_forwarded_to_tok_params():

26 """build_tok_params must forward return_token_offsets, defaulting to

27 False (zero behavioral change for existing callers) and coercing JSON

28 - null to False via the bool() guard.

29 - """

28 + null to False via the bool() guard."""

29 cfg = _model_config()

30

31 default = CompletionRequest(model="m", prompt="hi")

@@ -41,8 +40,7 @@ def test_completion_flag_forwarded_to_tok_params():

40

41 def test_chat_flag_forwarded_to_tok_params():

42 """Chat build_tok_params has its own (max_completion_tokens) branch, so

44 - its return_token_offsets forwarding is verified independently.

45 - """

43 + its return_token_offsets forwarding is verified independently."""

44 cfg = _model_config()

45 messages = [{"role": "user", "content": "hi"}]

46

@@ -66,8 +64,7 @@ def test_generate_request_token_offsets_default_none():

64

65 def test_generate_request_token_offsets_survive_json_round_trip():

66 """GenerateRequest crosses the disagg serialization boundary; the

69 - tuple[int, int] offsets must survive model_dump and re-validate.

70 - """

67 + tuple[int, int] offsets must survive model_dump and re-validate."""

68 req = GenerateRequest(

69 token_ids=[10, 20],

70 sampling_params=SamplingParams(),

tests/entrypoints/pooling/embed/test_cohere_online.py

+2 -4

@@ -111,8 +111,7 @@ def test_basic_embed(server: RemoteOpenAIServer, model_name: str):

111

112 def test_unsupported_input_type_rejected(server: RemoteOpenAIServer, model_name: str):

113 """An input_type not defined in the model's prompt config should be

114 - rejected with a 400 error.

115 - """

114 + rejected with a 400 error."""

115 body = {

116 "model": model_name,

117 "input_type": "nonexistent_type",

@@ -139,8 +138,7 @@ def test_omitted_input_type_accepted(server: RemoteOpenAIServer, model_name: str

138

139 def test_v1_v2_parity(server: RemoteOpenAIServer, model_name: str):

140 """v1 (OpenAI) and v2 (Cohere) endpoints should produce the same

142 - float embeddings for a generic model.

143 - """

141 + float embeddings for a generic model."""

142 texts = ["hello world"]

143 v2 = _cohere_embed(server, model_name, texts=texts, embedding_types=["float"])

144 v1 = _openai_embed(server, model_name, texts)

tests/entrypoints/pooling/embed/test_cohere_openai_parity.py

+1 -2

@@ -76,8 +76,7 @@ def test_single_text_parity(server: RemoteOpenAIServer):

76

77 def test_batch_parity(server: RemoteOpenAIServer):

78 """A batch of texts should produce equivalent embeddings via both APIs,

79 - in the same order.

80 - """

79 + in the same order."""

80 texts = [

81 "machine learning",

82 "deep learning",

tests/entrypoints/pooling/embed/test_io_processor.py

+1 -2

@@ -241,8 +241,7 @@ class TestResolveTruncation:

241

242 def test_truncate_none_with_max_tokens(self):

243 """truncate=NONE should NOT set truncate_prompt_tokens; the

244 - max_tokens limit is enforced separately via _check_max_tokens.

245 - """

244 + max_tokens limit is enforced separately via _check_max_tokens."""

245 req = self._make_request(truncate="NONE", max_tokens=10)

246 tokens, side = EmbedIOProcessor._resolve_cohere_truncation(req)

247 assert tokens is None

tests/entrypoints/pooling/embed/test_online_long_text.py

+3 -6

@@ -240,8 +240,7 @@ async def test_long_text_embedding_1500_chars(

240 client_with_chunked_processing: openai.AsyncOpenAI, model_name: str

241 ):

242 """Test embedding processing for ~1500 character long text

243 - (~1028 tokens, exceeding 512 token limit).

244 - """

243 + (~1028 tokens, exceeding 512 token limit)."""

244 # Verify text length

245 # Verify text has sufficient word count (approximately 1500 words)

246 word_count = len(LONG_TEXT_1500_WORDS.split())

@@ -289,8 +288,7 @@ async def test_long_text_embedding_2500_chars(

288 client_with_chunked_processing: openai.AsyncOpenAI, model_name: str

289 ):

290 """Test embedding processing for ~2500 character long text

292 - (~2048 tokens, requiring multiple chunks).

293 - """

291 + (~2048 tokens, requiring multiple chunks)."""

292 # Verify text length

293 # Verify text has sufficient word count (approximately 2500 words)

294 word_count = len(LONG_TEXT_2500_WORDS.split())

@@ -382,8 +380,7 @@ async def test_chunked_vs_normal_consistency(

380 client_with_chunked_processing: openai.AsyncOpenAI, model_name: str

381 ):

382 """Test consistency between chunked and

385 - normal processing (using short text).

386 - """

383 + normal processing (using short text)."""

384 # Use a short text within the 512 token limit

385 short_text = (

386 "Artificial intelligence technology is changing our world, "

tests/entrypoints/pooling/embed/test_protocol.py

+1 -2

@@ -1,8 +1,7 @@

1 # SPDX-License-Identifier: Apache-2.0

2 # SPDX-FileCopyrightText: Copyright contributors to the vLLM project

3 """Unit tests for Cohere embed protocol: build_typed_embeddings and its

4 -underlying packing helpers, plus Cohere-specific serving helpers.

5 -"""

4 +underlying packing helpers, plus Cohere-specific serving helpers."""

5

6 import struct

7

tests/entrypoints/scale_out/derender/test_derender.py

+1 -2

@@ -454,8 +454,7 @@ async def test_derender_completion_empty_generate_responses(client):

454 @pytest.mark.asyncio

455 async def test_derender_completion_logprobs(client):

456 """token_id:N placeholders in logprobs are resolved; CompletionLogProbs

457 - flat-list structure is returned with non-empty tokens and text_offsets.

458 - """

457 + flat-list structure is returned with non-empty tokens and text_offsets."""

458 gr1 = await _render_completion(client, "Hello world")

459 ids1 = gr1["token_ids"][:3]

460 token_id = ids1[0]

tests/entrypoints/scale_out/derender/test_derender_parity.py

+3 -6

@@ -121,8 +121,7 @@ async def _disagg(

121

122 def _tool_sig(response_choice: dict) -> list[tuple[str, dict]]:

123 """[(name, json normalized args)] so key ordering / whitespace don't

124 - cause false negatives.

125 - """

124 + cause false negatives."""

125 return [

126 (tc["function"]["name"], json.loads(tc["function"]["arguments"]))

127 for tc in (response_choice["message"].get("tool_calls") or [])

@@ -144,8 +143,7 @@ async def _run_parity_case(

143 client: httpx.AsyncClient, messages: list[dict], **extra

144 ) -> tuple[dict, dict]:

145 """Run the coupled request then feed its generated tokens into the

147 - disaggregated derender endpoint. Returns (coupled, disagg).

148 - """

146 + disaggregated derender endpoint. Returns (coupled, disagg)."""

147 coupled = await _coupled(client, messages, **extra)

148 ch = coupled["choices"][0]

149 chat_request = {"model": MODEL, "messages": messages, **extra}

@@ -205,8 +203,7 @@ async def test_parity_tool_call(client):

203 @pytest.mark.asyncio

204 async def test_parity_reasoning_and_tool_call(client):

205 """Combined reasoning + tool call parity means the highest drift risk

208 - since it exercises both parser branches on the same output.

209 - """

206 + since it exercises both parser branches on the same output."""

207 messages = [{"role": "user", "content": "What's the weather in Paris?"}]

208 coupled, disagg = await _run_parity_case(

209 client,

tests/entrypoints/scale_out/render/test_render.py

+4 -8

@@ -455,8 +455,7 @@ async def test_chat_completion_render_with_sampling_params(client):

455 @pytest.mark.asyncio

456 async def test_completion_render_emits_token_offsets(client):

457 """With return_token_offsets, /v1/completions/render returns per-token

458 - (start, end) char offsets aligned with token_ids.

459 - """

458 + (start, end) char offsets aligned with token_ids."""

459 prompt = "Hello, world."

460 response = await client.post(

461 "/v1/completions/render",

@@ -481,8 +480,7 @@ async def test_completion_render_emits_token_offsets(client):

480 @pytest.mark.asyncio

481 async def test_completion_render_default_no_token_offsets(client):

482 """Without the flag, token_offsets must be null (existing responses

484 - unchanged).

485 - """

483 + unchanged)."""

484 response = await client.post(

485 "/v1/completions/render",

486 json={

@@ -499,8 +497,7 @@ async def test_completion_render_default_no_token_offsets(client):

497 @pytest.mark.asyncio

498 async def test_chat_render_emits_token_offsets(client):

499 """With return_token_offsets, /v1/chat/completions/render returns

502 - per-token offsets relative to the templated prompt string.

503 - """

500 + per-token offsets relative to the templated prompt string."""

501 response = await client.post(

502 "/v1/chat/completions/render",

503 json={

@@ -674,8 +671,7 @@ async def test_chat_completion_render_assistant_tokens_mask_with_generation_tags

671 client,

672 ):

673 """With a ``{% generation %}``-enabled template, the mask marks assistant

677 - tokens and the masked tokens decode to the assistant content.

678 - """

674 + tokens and the masked tokens decode to the assistant content."""

675 response = await client.post(

676 "/v1/chat/completions/render",

677 json={

tests/entrypoints/scale_out/token_in_token_out/test_protocol.py

+3 -6

@@ -25,8 +25,7 @@ def _base_payload() -> dict:

25

26 def test_omitted_max_tokens_is_not_provided():

27 """Body without ``max_tokens`` must surface as 'not provided' so the

28 - server can apply its own default instead of the dataclass 16.

29 - """

28 + server can apply its own default instead of the dataclass 16."""

29 req = GenerateRequest.model_validate(_base_payload())

30 # SamplingParams' dataclass default leaks through the parsed instance —

31 # this is exactly the bug the server-side defaulting works around.

@@ -36,8 +35,7 @@ def test_omitted_max_tokens_is_not_provided():

35

36 def test_explicit_max_tokens_is_provided():

37 """Even when the client picks the same value as the dataclass default,

39 - it must register as explicitly set so the server won't override it.

40 - """

38 + it must register as explicitly set so the server won't override it."""

39 payload = _base_payload()

40 payload["sampling_params"] = {"max_tokens": 16}

41 req = GenerateRequest.model_validate(payload)

@@ -69,8 +67,7 @@ def test_json_roundtrip_preserves_provided_keys():

67 def test_internal_instance_construction_treats_all_as_provided():

68 """When internal callers build ``GenerateRequest`` from a pre-resolved

69 ``SamplingParams`` instance, every field is considered explicitly set

72 - so server-side defaulting can't clobber values resolved upstream.

73 - """

70 + so server-side defaulting can't clobber values resolved upstream."""

71 sp = SamplingParams(max_tokens=500, temperature=0.0)

72 req = GenerateRequest(token_ids=[1, 2, 3], sampling_params=sp)

73 assert req.is_sampling_param_provided("max_tokens")

tests/entrypoints/serve/exception_handling/test_error_sanitization.py

+2 -4

@@ -22,8 +22,7 @@ def test_sanitize_message():

22

23 class TestSanitizeMessageFilePaths:

24 """sanitize_message should also strip file paths and traceback

25 - frames, not just memory addresses - see #31683.

26 - """

25 + frames, not just memory addresses - see #31683."""

26

27 def test_strips_traceback_style_frame(self):

28 msg = (

@@ -45,8 +44,7 @@ class TestSanitizeMessageFilePaths:

44

45 def test_strips_single_parent_container_path(self):

46 """Regression: /app/server.py and /workspace/server.py (common in

48 - container deployments) were missed by the original {2,} quantifier.

49 - """

47 + container deployments) were missed by the original {2,} quantifier."""

48 assert "/app/" not in sanitize_message("Error in /app/server.py")

49 assert "/workspace/" not in sanitize_message("Error in /workspace/server.py")

50

tests/entrypoints/serve/exception_handling/test_validation_exception_handler.py

+5 -10

@@ -24,8 +24,7 @@ from vllm.entrypoints.serve.exception_handling.handlers.validation import (

24

25 def _fake_request(log_error_stack: bool = False) -> SimpleNamespace:

26 """Minimal stand-in for a FastAPI Request - just enough for the

27 - handler to read req.app.state.args.log_error_stack.

28 - """

27 + handler to read req.app.state.args.log_error_stack."""

28 return SimpleNamespace(

29 app=SimpleNamespace(

30 state=SimpleNamespace(args=SimpleNamespace(log_error_stack=log_error_stack))

@@ -36,8 +35,7 @@ def _fake_request(log_error_stack: bool = False) -> SimpleNamespace:

35

36 class TestValidationErrorParamFallback:

37 """Ensure `param` falls back to the Pydantic error's `loc` when no

39 - custom VLLMValidationError context is present.

40 - """

38 + custom VLLMValidationError context is present."""

39

40 @pytest.mark.parametrize(

41 ("error_type", "msg"),

@@ -61,8 +59,7 @@ class TestValidationErrorParamFallback:

59 async def test_param_fallback_does_not_crash_on_non_dict_error(self):

60 """Schemathesis fuzzing found that errors[0] isn't always a dict.

61 The fallback must not crash in that case - it should just leave

64 - param as None instead of raising.

65 - """

62 + param as None instead of raising."""

63 exc = RequestValidationError(["some unexpected non-dict error"])

64

65 response = await validation_exception_handler(_fake_request(), exc)

@@ -100,8 +97,7 @@ class TestCleanLocForParam:

97

98 def test_all_internal_falls_back_to_raw_join(self):

99 """If every loc segment looks internal, fall back to the raw

103 - dot-join rather than returning an empty string.

104 - """

100 + dot-join rather than returning an empty string."""

101 loc = ("function-wrap[__log_extra_fields__()]",)

102 assert clean_loc_for_param(loc) == "function-wrap[__log_extra_fields__()]"

103

@@ -109,8 +105,7 @@ class TestCleanLocForParam:

105 class TestValidationErrorDoesNotLeakServerPaths:

106 """FastAPI sets endpoint_file/line/function/path on the exception

107 during routing, not in the constructor - so we set them manually

112 - here to reproduce the real leak. See #31683.

113 - """

108 + here to reproduce the real leak. See #31683."""

109

110 @pytest.mark.asyncio

111 async def test_handler_strips_endpoint_file_context(self):

tests/entrypoints/serve/lora/test_lora_adapters.py

+1 -2

@@ -245,8 +245,7 @@ async def test_multiple_lora_adapters(

245 client: openai.AsyncOpenAI, tmp_path, qwen3_lora_files

246 ):

247 """Validate that many loras can be dynamically registered and inferenced

248 - with concurrently

249 - """

248 + with concurrently"""

249

250 # This test file configures the server with --max-cpu-loras=2 and this test

251 # will concurrently load 10 adapters, so it should flex the LRU cache

tests/entrypoints/serve/middleware/test_authentication_middleware.py

+3 -6

@@ -72,8 +72,7 @@ class MockModelConfig:

72 @pytest.fixture(params=get_args(SupportedTask))

73 def task_routes(request, monkeypatch) -> tuple[str, list[tuple[str, list[str]]]]:

74 """For each supported task, build an app with only that task's routers,

75 - extract all routes, and return the task name and routes.

76 - """

75 + extract all routes, and return the task name and routes."""

76 task = request.param

77 # Enable development mode to register all routes (including dev-only routes).

78 monkeypatch.setenv("VLLM_SERVER_DEV_MODE", "1")

@@ -99,8 +98,7 @@ def task_routes(request, monkeypatch) -> tuple[str, list[tuple[str, list[str]]]]

98

99 def test_auto_discovered_protected_routes_require_auth(task_routes):

100 """For every auto-discovered route that starts with a guarded prefix,

102 - verify that authentication is enforced.

103 - """

101 + verify that authentication is enforced."""

102 task, routes = task_routes

103 app = _create_app_with_mock_routes(routes)

104 client = TestClient(app)

@@ -134,8 +132,7 @@ def test_auto_discovered_protected_routes_require_auth(task_routes):

132

133 def test_auto_discovered_unprotected_routes_no_auth(task_routes):

134 """For every auto-discovered route that does NOT start with a guarded

137 - prefix, verify that no authentication is required.

138 - """

135 + prefix, verify that no authentication is required."""

136 task, routes = task_routes

137 app = _create_app_with_mock_routes(routes)

138 client = TestClient(app)

tests/entrypoints/serve/sagemaker/test_sagemaker_handler_overrides.py

+6 -12

@@ -114,8 +114,7 @@ class TestHandlerOverrideIntegration:

114

115 def test_customer_script_functions_auto_loaded(self, monkeypatch, tmp_path):

116 """Test customer scenario: script functions automatically override

117 - framework defaults.

118 - """

117 + framework defaults."""

118 try:

119 from model_hosting_container_standards.sagemaker.config import (

120 SageMakerEnvVars,

@@ -176,8 +175,7 @@ async def custom_sagemaker_invocation_handler(request: Request):

175

176 def test_customer_decorator_usage(self, monkeypatch, tmp_path):

177 """Test customer scenario: using @custom_ping_handler and

179 - @custom_invocation_handler decorators.

180 - """

178 + @custom_invocation_handler decorators."""

179 try:

180 from model_hosting_container_standards.sagemaker.config import (

181 SageMakerEnvVars,

@@ -233,8 +231,7 @@ async def my_invoke(request: Request):

231

232 def test_handler_priority_order(self, monkeypatch, tmp_path):

233 """Test priority: @custom_ping_handler/@custom_invocation_handler

236 - decorators vs script functions.

237 - """

234 + decorators vs script functions."""

235 try:

236 from model_hosting_container_standards.sagemaker.config import (

237 SageMakerEnvVars,

@@ -307,8 +304,7 @@ async def custom_sagemaker_invocation_handler(request: Request):

304

305 def test_environment_variable_script_loading(self, monkeypatch, tmp_path):

306 """Test that environment variables correctly specify script location

310 - and loading.

311 - """

307 + and loading."""

308 try:

309 from model_hosting_container_standards.sagemaker.config import (

310 SageMakerEnvVars,

@@ -420,8 +416,7 @@ async def custom_sagemaker_invocation_handler(request: Request):

416

417 def test_handler_env_var_override(self, monkeypatch, tmp_path):

418 """Test CUSTOM_FASTAPI_PING_HANDLER and CUSTOM_FASTAPI_INVOCATION_HANDLER

423 - environment variable overrides.

424 - """

419 + environment variable overrides."""

420 try:

421 from model_hosting_container_standards.common.fastapi.config import (

422 FastAPIEnvVars,

@@ -515,8 +510,7 @@ async def custom_sagemaker_invocation_handler(request: Request):

510

511 def test_env_var_priority_over_decorator_and_script(self, monkeypatch, tmp_path):

512 """Test that environment variables have highest priority over decorators

518 - and script functions for both ping and invocation handlers.

519 - """

513 + and script functions for both ping and invocation handlers."""

514 try:

515 from model_hosting_container_standards.common.fastapi.config import (

516 FastAPIEnvVars,

tests/entrypoints/serve/tokenize/test_tokenize_then_chat_vlm.py

+1 -2

@@ -51,8 +51,7 @@ async def test_tokenize_then_chat_completion_with_image(

51 local_asset_server,

52 ):

53 """Tokenize a multimodal message, then send the same message to chat

54 - completions. The chat completion must succeed (not 500).

55 - """

54 + completions. The chat completion must succeed (not 500)."""

55 image_url = local_asset_server.url_for("stop_sign.jpg")

56 messages = [

57 {

tests/entrypoints/serve/utils/test_api_utils.py

+4 -8

@@ -36,13 +36,11 @@ def test_should_include_usage_force_enables_continuous_usage(stream_options, exp

36 class TestGetMaxTokens:

37 """Tests for get_max_tokens() to ensure generation_config's max_tokens

38 acts as a default when from model author, and as a ceiling when

39 - explicitly set by the user.

40 - """

39 + explicitly set by the user."""

40

41 def test_default_sampling_params_used_when_no_request_max_tokens(self):

42 """When user doesn't specify max_tokens, generation_config default

44 - should apply.

45 - """

43 + should apply."""

44 result = get_max_tokens(

45 max_model_len=24000,

46 max_tokens=None,

@@ -53,8 +51,7 @@ class TestGetMaxTokens:

51

52 def test_request_max_tokens_not_capped_by_default_sampling_params(self):

53 """When user specifies max_tokens in request, model author's

56 - generation_config max_tokens must NOT cap it (fixes #34005).

57 - """

54 + generation_config max_tokens must NOT cap it (fixes #34005)."""

55 result = get_max_tokens(

56 max_model_len=24000,

57 max_tokens=5000,

@@ -97,8 +94,7 @@ class TestGetMaxTokens:

94

95 def test_request_max_tokens_smaller_than_default(self):

96 """When user explicitly requests fewer tokens than gen_config default,

100 - that should be respected.

101 - """

97 + that should be respected."""

98 result = get_max_tokens(

99 max_model_len=24000,

100 max_tokens=512,

tests/entrypoints/serve/utils/test_request_logger.py

+1 -2

@@ -239,8 +239,7 @@ def test_request_logger_log_outputs_integration():

239

240 def test_streaming_complete_logs_full_text_content():

241 """Test that streaming complete logging includes

242 - full accumulated text, not just token count.

243 - """

242 + full accumulated text, not just token count."""

243 mock_logger = MagicMock()

244

245 with patch("vllm.entrypoints.serve.utils.request_logger.logger", mock_logger):

tests/entrypoints/serve/utils/test_sse_keep_alive.py

+8 -16

@@ -17,8 +17,7 @@ from vllm.entrypoints.serve.utils.sse_keep_alive import (

17 def test_disabled_returns_same_generator():

18 """For 0, negative, and NaN intervals the exact same generator object must

19 be returned (identity), proving no wrapper or per-chunk overhead when

20 - disabled.

21 - """

20 + disabled."""

21

22 async def gen():

23 yield "x"

@@ -34,8 +33,7 @@ def test_disabled_returns_same_generator():

33 @pytest.mark.asyncio

34 async def test_keep_alive_during_silent_start():

35 """A silent start (queue wait / prefill) emits keep-alive comments before

37 - the first real chunk.

38 - """

36 + the first real chunk."""

37 release = asyncio.Event()

38

39 async def gen():

@@ -57,8 +55,7 @@ async def test_keep_alive_during_silent_start():

55 @pytest.mark.asyncio

56 async def test_no_keep_alive_when_streaming_is_fast():

57 """If chunks arrive faster than the interval, no comment is emitted and all

60 - data is forwarded.

61 - """

58 + data is forwarded."""

59

60 async def gen():

61 for chunk in ["a", "b", "c"]:

@@ -72,8 +69,7 @@ async def test_no_keep_alive_when_streaming_is_fast():

69 @pytest.mark.asyncio

70 async def test_keep_alive_between_chunks_does_not_drop_data():

71 """Keep-alives interleaved with real chunks must never drop or reorder

75 - data.

76 - """

72 + data."""

73 release_b = asyncio.Event()

74 release_c = asyncio.Event()

75

@@ -139,8 +135,7 @@ async def test_upstream_cancelled_error_propagates():

135 @pytest.mark.asyncio

136 async def test_aclose_closes_upstream():

137 """Closing a wrapper that is mid-stream must complete promptly and close

142 - the upstream generator (regression for the earlier leaked-task hang).

143 - """

138 + the upstream generator (regression for the earlier leaked-task hang)."""

139 finalized = asyncio.Event()

140

141 async def gen():

@@ -159,8 +154,7 @@ async def test_aclose_closes_upstream():

154 @pytest.mark.asyncio

155 async def test_client_disconnect_cancels_upstream():

156 """Cancelling the consumer mid-stream must cancel the pending anext and

162 - close the upstream generator so its cleanup runs.

163 - """

157 + close the upstream generator so its cleanup runs."""

158 finalized = asyncio.Event()

159

160 async def gen():

@@ -187,8 +181,7 @@ async def test_client_disconnect_cancels_upstream():

181 @pytest.mark.asyncio

182 async def test_streaming_response_emits_comment():

183 """A keep-alive comment must reach the client through the real

190 - StreamingResponse used by the routers.

191 - """

184 + StreamingResponse used by the routers."""

185 release = asyncio.Event()

186

187 async def gen():

@@ -227,8 +220,7 @@ async def test_streaming_response_emits_comment():

220 async def test_streaming_response_disconnect_closes_upstream():

221 """On the ASGI 2.3 disconnect path used by the pinned Uvicorn, Starlette

222 cancels streaming and the wrapper's finally must close the upstream

230 - generator.

231 - """

223 + generator."""

224 finalized = asyncio.Event()

225

226 async def gen():

tests/entrypoints/speech_to_text/realtime/test_realtime_validation.py

+1 -2

@@ -295,8 +295,7 @@ async def test_session_update_invalid_model_returns_error(model_name):

295 @pytest.mark.parametrize("model_name", [MODEL_NAME])

296 async def test_commit_without_session_update_returns_error(model_name):

297 """Test that committing before validating the model returns an error

298 - and does not fall through to processing.

299 - """

298 + and does not fall through to processing."""

299 server_args = ["--enforce-eager", "--max-model-len", "2048"]

300

301 if model_name.startswith("mistralai"):

tests/entrypoints/unit_tests/test_context.py

+7 -14

@@ -51,8 +51,7 @@ async def generate_mock_outputs(

51 num_turns, prompt_token_counts, output_token_counts, cached_token_counts=None

52 ):

53 """Generate a sequence of mock RequestOutput objects to simulate multiple

54 - turns.

55 - """

54 + turns."""

55 if cached_token_counts is None:

56 cached_token_counts = [0] * num_turns

57

@@ -271,8 +270,7 @@ def test_reasoning_tokens_counting():

270

271 def test_preamble_tokens_not_counted_as_reasoning():

272 """Preambles (commentary with no recipient) are visible user text,

274 - not hidden reasoning. They must NOT inflate num_reasoning_tokens.

275 - """

273 + not hidden reasoning. They must NOT inflate num_reasoning_tokens."""

274 context, parser = make_harmony_context()

275 parser.enqueue_chunk_result(reasoning_token_count=0)

276

@@ -289,8 +287,7 @@ def test_preamble_tokens_not_counted_as_reasoning():

287

288 def test_commentary_with_recipient_counted_as_reasoning():

289 """Commentary directed at a tool (recipient != None) is hidden from

292 - the user, so it should still count as reasoning tokens.

293 - """

290 + the user, so it should still count as reasoning tokens."""

291 context, parser = make_harmony_context()

292 parser.enqueue_chunk_result(reasoning_token_count=3)

293

@@ -349,8 +346,7 @@ async def test_single_turn_no_tool_output():

346 @pytest.mark.asyncio

347 async def test_negative_tool_tokens_edge_case():

348 """Test edge case where calculation could result in negative tool

352 - tokens. We should log an error and clamp the value to 0.

353 - """

349 + tokens. We should log an error and clamp the value to 0."""

350 # Use patch to check if logger.error was called

351 with patch("vllm.entrypoints.openai.responses.context.logger.error") as mock_log:

352 context, _ = make_harmony_context(available_tools=["browser"])

@@ -690,8 +686,7 @@ def create_simple_context_output(

686 finished=True,

687 ):

688 """Helper to create a RequestOutput with customizable text for

693 - SimpleContext tests.

694 - """

689 + SimpleContext tests."""

690 if token_ids is None:

691 token_ids = []

692 return RequestOutput(

@@ -875,8 +870,7 @@ def test_simple_context_final_output():

870

871 def test_simple_context_output_messages_empty_text_with_tokens():

872 """output_messages should be returned when tokens exist even if text is

878 - empty (e.g. special tokens).

879 - """

873 + empty (e.g. special tokens)."""

874 context = SimpleContext()

875 context.append_output(

876 create_simple_context_output(

@@ -894,8 +888,7 @@ def test_simple_context_output_messages_empty_text_with_tokens():

888

889 def test_simple_context_output_messages_no_mutation():

890 """Each call to output_messages returns a fresh list; callers can't

897 - corrupt internal state.

898 - """

891 + corrupt internal state."""

892 context = SimpleContext()

893 context.append_output(

894 create_simple_context_output(

tests/entrypoints/weight_transfer/test_weight_transfer_llm.py

+1 -2

@@ -121,8 +121,7 @@ def test_get_world_size_tp1(vllm_runner):

121 @create_new_process_for_each_test()

122 def test_init_weight_transfer_engine_calls_engine(vllm_runner):

123 """Test that init_weight_transfer_engine calls the engine's

124 - init_transfer_engine method.

125 - """

124 + init_transfer_engine method."""

125 if torch.accelerator.device_count() < 1:

126 pytest.skip("Need at least 1 GPU for this test")

127

tests/fusion/test_quant_activation_contract.py

+1 -2

@@ -86,8 +86,7 @@ def _all_kernel_classes() -> list[type]:

86

87 def _probe(cls: type):

88 """A bare kernel instance with a plausible config, so input_quant_key()

89 - can be queried without the hardware-gated constructor.

90 - """

89 + can be queried without the hardware-gated constructor."""

90 obj = cls.__new__(cls) # type: ignore[call-overload]

91 if issubclass(cls, NvFp4LinearKernel):

92 obj.config = NvFp4LinearLayerConfig()

tests/kernels/attention/test_amx_mla.py

+3 -6

@@ -247,8 +247,7 @@ def test_extend_attention_cpu_matches_reference(prefix_lens, extend_lens):

247

248 class _FakeLinear:

249 """Minimal duck-typed stand-in for kv_b_proj: only .weight and

250 - .quant_method are read by get_and_maybe_dequant_weights().

251 - """

250 + .quant_method are read by get_and_maybe_dequant_weights()."""

251

252 def __init__(self, weight: torch.Tensor):

253 self.weight = weight

@@ -319,8 +318,7 @@ class _FakeAttnMetadata:

318 def test_amx_mla_impl_forward_mqa_matches_reference(default_vllm_config):

319 """forward_mqa receives already-absorbed Q (as MLAAttention.forward_impl

320 would produce via layer.W_UK_T) and must reproduce plain per-head MLA

322 - decode attention through W_UK/W_UV.

323 - """

321 + decode attention through W_UK/W_UV."""

322 set_random_seed(1)

323 num_heads, qk_nope_head_dim, v_head_dim, kv_lora_rank = 4, 32, 32, 64

324 impl, kv_b_weight = _make_amx_mla_impl(

@@ -386,8 +384,7 @@ def test_amx_mla_impl_forward_mqa_matches_reference(default_vllm_config):

384 def test_amx_mla_impl_forward_mha_matches_reference(default_vllm_config):

385 """forward_mha receives raw unabsorbed Q and must attend correctly for a

386 prefill batch with a mix of fresh and continued (cached-prefix)

389 - sequences.

390 - """

387 + sequences."""

388 set_random_seed(2)

389 num_heads, qk_nope_head_dim, v_head_dim, kv_lora_rank = 4, 32, 32, 64

390 impl, kv_b_weight = _make_amx_mla_impl(

tests/kernels/attention/test_attention_selector.py

+2 -4

@@ -616,8 +616,7 @@ def test_non_causal_autoselect_backend():

616 )

617 def test_flash_attn_rejects_unhandled_kv_cache_dtypes(kv_cache_dtype: str):

618 """FlashAttentionBackend must not claim support for kv_cache dtypes

619 - that it cannot handle.

620 - """

619 + that it cannot handle."""

620 from vllm.v1.attention.backends.flash_attn import FlashAttentionBackend

621

622 assert not FlashAttentionBackend.supports_kv_cache_dtype(kv_cache_dtype)

@@ -628,8 +627,7 @@ def test_flash_attn_accepts_handled_fp8_variants(

627 kv_cache_dtype: str, monkeypatch: pytest.MonkeyPatch

628 ):

629 """FlashAttentionBackend must accept the two fp8 dtypes it can actually

631 - handle: 'fp8' (alias for fp8_e4m3fn) and 'fp8_e4m3'.

632 - """

630 + handle: 'fp8' (alias for fp8_e4m3fn) and 'fp8_e4m3'."""

631 import vllm.v1.attention.backends.fa_utils as fa_utils_mod

632 from vllm.v1.attention.backends.flash_attn import FlashAttentionBackend

633

tests/kernels/attention/test_cutlass_mla_decode.py

+1 -2

@@ -225,8 +225,7 @@ def test_cutlass_mla_decode_cross_layer_view():

225 """The kernel must read the cache's page-dim stride instead of assuming

226 pages are packed back-to-back. A per-layer view into a cross-layer

227 (block-major) cache has stride(0) inflated by num_layers; outputs must

228 - match a contiguous cache holding the same data exactly.

229 - """

228 + match a contiguous cache holding the same data exactly."""

229 device = torch.device("cuda:0")

230 torch.set_default_dtype(torch.bfloat16)

231 torch.set_default_device(device)

tests/kernels/attention/test_flashinfer.py

+1 -2

@@ -162,8 +162,7 @@ def test_fast_decode_plan_importable() -> None:

162 @torch.inference_mode

163 def test_fast_plan_decode_warmup_uses_full_plan(dtype: torch.dtype) -> None:

164 """On the first call fast_plan_decode must route through self.plan() and

165 - flip vllm_first_call to False on the wrapper object.

166 - """

165 + flip vllm_first_call to False on the wrapper object."""

166 from unittest.mock import patch

167

168 from vllm.v1.attention.backends.flashinfer import fast_plan_decode

tests/kernels/attention/test_flashinfer_trtllm_attention.py

+1 -2

@@ -134,8 +134,7 @@ def make_quantized_kv_cache(

134 head_size: int,

135 ) -> tuple:

136 """Quantize kv_cache based on dtype. Returns (kv_cache, kv_cache_sf,

137 - kv_scale, ref_kv_cache, is_nvfp4_kv).

138 - """

137 + kv_scale, ref_kv_cache, is_nvfp4_kv)."""

138 is_nvfp4_kv = kv_quant_dtype == FP4_DTYPE

139 if is_nvfp4_kv:

140 data, scales, kv_scale, ref = make_nvfp4_kv_cache(

tests/kernels/attention/test_kimi_k3_mla_fused_epilogue.py

+1 -2

@@ -73,8 +73,7 @@ def _strided_context_inputs(

73 ) -> tuple[torch.Tensor, torch.Tensor, torch.Tensor]:

74 """The real prefill-context layouts: ``k_nope``/``v`` as the two strided

75 halves of one kv_b_proj output and ``k_pe`` as a column slice of the gather

76 - workspace (left in the fp8 cache layout for a plain fp8 cache).

77 - """

76 + workspace (left in the fp8 cache layout for a plain fp8 cache)."""

77 kv_nope = _randn(num_tokens, num_heads, 256, dtype=dtype)

78 k_nope, v = kv_nope.split((128, 128), dim=-1)

79 workspace = _randn(num_tokens, 576, dtype=dtype)

tests/kernels/attention/test_merge_attn_states.py

+1 -2

@@ -89,8 +89,7 @@ all_case_info: list[tuple] = []

89 @pytest.mark.parametrize("output_dtype", [torch.float32, torch.half, torch.bfloat16])

90 def test_merge_attn_states_both_empty(merge_fn, output_dtype) -> None:

91 """When a token is empty on both sides (both LSE -inf), the 0/0 softmax

92 - scales must not surface as NaN in the merged output.

93 - """

92 + scales must not surface as NaN in the merged output."""

93 num_tokens, num_heads, head_size = 6, 8, 128

94 prefix_output = torch.zeros(

95 num_tokens, num_heads, head_size, device=DEVICE, dtype=output_dtype

tests/kernels/attention/test_minimax_m3.py

+14 -28

@@ -45,8 +45,7 @@ def kv_layout(request) -> KVCacheLayout:

45

46 def _layer_stride_order(layout: KVCacheLayout, ndim: int) -> tuple[int, ...]:

47 """Per-layer physical stride order for the given layout; the 3-dim

48 - indexer side cache (H=1) is contiguous, so identity.

49 - """

48 + indexer side cache (H=1) is contiguous, so identity."""

49 if ndim == 3:

50 return (0, 1, 2)

51 stride_order = layout.layer_view_order

@@ -66,8 +65,7 @@ def _main_spec() -> FullAttentionSpec:

65

66 def _main_kv_logical_shape(num_pages: int) -> tuple[int, ...]:

67 """Standardized per-layer logical shape [B, H, N, C] for the main cache,

69 - derived the same way the production allocator does.

70 - """

68 + derived the same way the production allocator does."""

69 shape_bytes = compute_layer_kv_cache_shape_bytes(_main_spec(), num_pages)

70 return (*shape_bytes[:-1], shape_bytes[-1] // DTYPE.itemsize)

71

@@ -77,8 +75,7 @@ def _allocate_main_kv_via_contract(

75 ) -> torch.Tensor:

76 """Build the main KV cache exactly as the production allocator does for the

77 given layout: allocate the physical (permuted) tensor, then expose the

80 - inverse-permuted logical [B, H, N, C] view the backend sees.

81 - """

78 + inverse-permuted logical [B, H, N, C] view the backend sees."""

79 logical_shape = _main_kv_logical_shape(num_pages)

80 stride_order = _layer_stride_order(layout, len(logical_shape))

81 physical_shape = tuple(logical_shape[i] for i in stride_order)

@@ -1379,8 +1376,7 @@ def test_amd_decode_index_topk_end_to_end(

1376 @pytest.mark.parametrize("num_idx_heads", [1, 4])

1377 def test_decode_index_topk_fp8(num_idx_heads: int):

1378 """The standalone Triton path must score FP8 inputs in FP32 so its top-k

1382 - matches a reference computed from the dequantized FP8 values.

1383 - """

1379 + matches a reference computed from the dequantized FP8 values."""

1380 torch.manual_seed(0)

1381 topk, init_blocks, local_blocks, head_dim = 8, 0, 1, 128

1382 decode_query_len = 1

@@ -1653,8 +1649,7 @@ def test_prefill_sparse_attention_correctness(

1649 def test_main_cache_layout_contract():

1650 """The standardized per-layer logical shape is [B, H, N, C] with packed

1651 K/V content, and the legacy layout aliases resolve to the expected

1656 - per-layer stride orders.

1657 - """

1652 + per-layer stride orders."""

1653 nb, bs, h, d = 7, BLOCK_SIZE, NUM_KV_HEADS, HEAD_DIM

1654 logical = _main_kv_logical_shape(nb)

1655 assert logical == (nb, h, bs, 2 * d)

@@ -1778,8 +1773,7 @@ def test_indexer_cache_squeezes_to_contiguous_3d():

1773 """The indexer side cache is standardized 4D with H=1: under both layouts

1774 the allocator's logical view stays contiguous and squeezes (as

1775 `MiniMaxM3IndexerCache.bind_kv_cache` does) to the 3-dim

1781 - [num_blocks, block_size, head_dim] cache the kernels consume.

1782 - """

1776 + [num_blocks, block_size, head_dim] cache the kernels consume."""

1777 nb = 5

1778 ispec = MLAAttentionSpec(

1779 block_size=BLOCK_SIZE, num_kv_heads=1, head_size=HEAD_DIM, dtype=DTYPE

@@ -1799,8 +1793,7 @@ def test_indexer_cache_squeezes_to_contiguous_3d():

1793

1794 def test_hnd_allocation_is_packed_head_major():

1795 """Under HND the backend-visible logical view is the packed head-major

1802 - physical allocation.

1803 - """

1796 + physical allocation."""

1797 nb, bs, h, d = 4, BLOCK_SIZE, NUM_KV_HEADS, HEAD_DIM

1798 logical = _main_kv_logical_shape(nb)

1799 stride_order = KVCacheLayout.LBHNC.layer_view_order

@@ -1828,8 +1821,7 @@ def test_hnd_allocation_is_packed_head_major():

1821 def test_main_cache_is_block_first_and_unpadded():

1822 """The allocator's contiguous-view branch (not the padded-strided branch)

1823 is used for the main GQA cache: its spec is unpadded and the physical

1831 - layout keeps num_blocks as the first dimension under both layouts.

1832 - """

1824 + layout keeps num_blocks as the first dimension under both layouts."""

1825 spec = _main_spec()

1826 # Unpadded -> allocator uses kv_tensor.view(...) rather than as_strided().

1827 assert spec.page_size_padded is None

@@ -1851,8 +1843,7 @@ def _build_decode_inputs(

1843 ):

1844 """Shared decode setup: uniform query tokens per request, a non-identity

1845 block table, and topk indices selecting the current block plus older causal

1854 - blocks for each query token.

1855 - """

1846 + blocks for each query token."""

1847 active_batch = len(seq_lens_list)

1848 batch = active_batch + num_padded_reqs

1849 pages_per_req = [(s + BLOCK_SIZE - 1) // BLOCK_SIZE for s in seq_lens_list]

@@ -2188,8 +2179,7 @@ def test_decode_sparse_attention_correctness(

2179 ):

2180 """Decode (split-K) parity under both layouts: this is the only coverage of

2181 the decode-site cache feed, and the strided HND case fails if the kernel

2191 - ignores the cache strides.

2192 - """

2182 + ignores the cache strides."""

2183 torch.manual_seed(0)

2184 q, block_table, seq_lens, topk_idx, num_pages = _build_decode_inputs(

2185 seq_lens_list, decode_query_len, num_padded_reqs

@@ -2239,8 +2229,7 @@ def test_decode_wrong_layout_breaks_parity():

2229 already contiguous-NHD (i.e. skipping the allocator's inverse permute)

2230 reorders the K/V content, so the decode output no longer matches the

2231 reference computed on the correct logical view. The mislabeled tensor keeps

2242 - the same shape as the correct view, so the kernel stays in bounds.

2243 - """

2232 + the same shape as the correct view, so the kernel stays in bounds."""

2233 torch.manual_seed(0)

2234 seq_lens_list = (130, 257)

2235 q, block_table, seq_lens, topk_idx, num_pages = _build_decode_inputs(seq_lens_list)

@@ -2275,8 +2264,7 @@ def test_decode_wrong_layout_breaks_parity():

2264 def test_main_cache_byte_identical_through_production_allocator():

2265 """AC-2: drive the real allocator (`create_kv_cache_views`) for the M3 main

2266 `FullAttentionSpec` under HND and assert the backend-visible view has the

2278 - same shape, stride, and storage offset as the packed-HND allocation.

2279 - """

2267 + same shape, stride, and storage offset as the packed-HND allocation."""

2268 nb = 4

2269 spec = _main_spec()

2270 raw = torch.zeros(nb * spec.page_size_bytes, dtype=torch.int8)

@@ -2291,8 +2279,7 @@ def test_main_cache_byte_identical_through_production_allocator():

2279 def test_padded_main_cache_is_flagged():

2280 """AC-2.1 negative: the M3 main cache relies on the allocator's

2281 contiguous-view branch (`page_size_padded is None`). A spec that sets

2294 - `page_size_padded` is explicitly flagged rather than silently wrong-strided.

2295 - """

2282 + `page_size_padded` is explicitly flagged rather than silently wrong-strided."""

2283

2284 def _require_unpadded_block_first(spec, stride_order):

2285 inv_order = [stride_order.index(i) for i in range(len(stride_order))]

@@ -2329,8 +2316,7 @@ def test_padded_main_cache_is_flagged():

2316 def test_reshape_and_cache_flash_write_persists(kv_layout: KVCacheLayout):

2317 """AC-5 write path: the `reshape_and_cache_flash` write site now consumes

2318 packed-content K/V split views. Writing through those views must persist

2332 - into the bound storage under both layouts.

2333 - """

2319 + into the bound storage under both layouts."""

2320 torch.manual_seed(0)

2321 num_pages = 4

2322 kv_cache = _allocate_main_kv_via_contract(num_pages, kv_layout)

tests/kernels/attention/test_mla_cross_layer_kernel_equivalence.py

+8 -16

@@ -23,8 +23,7 @@ pytestmark = pytest.mark.skipif(

23 def test_concat_and_cache_mla_into_unified_slot_view():

24 """concat_and_cache_mla must write correctly into a per-layer view whose

25 block stride is the full unified slot (block-major), with zero bleed into

26 - the other layers' segments of the same slot.

27 - """

26 + the other layers' segments of the same slot."""

27 from vllm import _custom_ops as ops

28

29 torch.manual_seed(0)

@@ -86,8 +85,7 @@ def test_concat_and_cache_mla_into_unified_slot_view():

85 def test_flashmla_dense_decode_unified_slot_view():

86 """FlashMLA dense decode (FLASHMLA backend, e.g. Kimi-K2-style dense MLA

87 on Hopper) must read a unified-slot block-major view bit-identically to a

89 - contiguous per-layer cache.

90 - """

88 + contiguous per-layer cache."""

89 import vllm.v1.attention.ops.flashmla as fm

90

91 ok, reason = fm.is_flashmla_dense_supported()

@@ -152,8 +150,7 @@ def test_flashmla_dense_decode_unified_slot_view():

150 def test_flashinfer_mla_dense_decode_unified_slot_view():

151 """FlashInfer MLA dense decode must read a unified-slot block-major view

152 (inflated stride(0), non-zero storage offset) bit-identically to a

155 - contiguous per-layer cache.

156 - """

153 + contiguous per-layer cache."""

154 try:

155 from flashinfer.decode import trtllm_batch_decode_with_kv_cache_mla

156 except ImportError:

@@ -224,8 +221,7 @@ def test_flashinfer_mla_dense_decode_unified_slot_view():

221 def test_flashmla_fp8_sparse_decode_unified_slot_view(next_n: int):

222 """FlashMLA fp8 sparse decode (DeepSeek V3.2/V4 DSA path) must read a

223 unified-slot block-major view bit-identically to a contiguous fp8_ds_mla

227 - cache, with finite nonzero output.

228 - """

224 + cache, with finite nonzero output."""

225 import vllm.v1.attention.ops.flashmla as fm

226

227 ok, reason = fm.is_flashmla_sparse_supported()

@@ -314,8 +310,7 @@ def test_flashmla_fp8_sparse_decode_unified_slot_view(next_n: int):

310 def test_indexer_k_quant_and_cache_into_unified_slot_view():

311 """indexer_k_quant_and_cache (DeepSeek V3.2/V4 DSA indexer K write) must

312 write correctly into a per-layer view whose block stride is the full

317 - unified slot, with zero bleed into the other layers' segments.

318 - """

313 + unified slot, with zero bleed into the other layers' segments."""

314 from vllm import _custom_ops as ops

315

316 torch.manual_seed(0)

@@ -360,8 +355,7 @@ def test_indexer_k_quant_and_cache_into_unified_slot_view():

355

356 def test_flashattn_mla_dense_decode_unified_slot_view():

357 """FA3 decode (FLASH_ATTN_MLA backend) must read a unified-slot

363 - block-major view bit-identically to a contiguous per-layer cache.

364 - """

358 + block-major view bit-identically to a contiguous per-layer cache."""

359 try:

360 from vllm.vllm_flash_attn import flash_attn_varlen_func

361 except ImportError:

@@ -434,8 +428,7 @@ def test_flashattn_mla_dense_decode_unified_slot_view():

428 def test_flashmla_dense_fp8_decode_unified_slot_view():

429 """FlashMLA dense fp8 decode (FLASHMLA backend with quantized KV cache)

430 must read a unified-slot block-major view bit-identically to a contiguous

437 - per-layer fp8 cache.

438 - """

431 + per-layer fp8 cache."""

432 import vllm.v1.attention.ops.flashmla as fm

433

434 ok, reason = fm.is_flashmla_dense_supported()

@@ -506,8 +499,7 @@ def test_flashmla_dense_fp8_decode_unified_slot_view():

499 def test_flashinfer_mla_dense_fp8_decode_unified_slot_view():

500 """FlashInfer MLA dense decode with an fp8 KV cache must read a

501 unified-slot block-major view bit-identically to a contiguous per-layer

509 - cache.

510 - """

502 + cache."""

503 try:

504 from flashinfer.decode import trtllm_batch_decode_with_kv_cache_mla

505 except ImportError:

tests/kernels/attention/test_rocm_aiter_fa.py

+7 -14

@@ -283,8 +283,7 @@ def _run_direct_flash_attn_case(

283 # Backend contract tests --------------------------------------------------

284 def test_aiter_mha_backend_contract():

285 """The ROCm backend advertises the dtypes, shapes, and attention types it

286 - is designed to handle.

287 - """

286 + is designed to handle."""

287 from vllm.v1.attention.backend import AttentionType

288 from vllm.v1.attention.backends.rocm_aiter_fa import AiterFlashAttentionBackend

289

@@ -312,8 +311,7 @@ def test_aiter_mha_backend_contract():

311

312 def test_aiter_mha_backend_validates_kv_cache_block_size():

313 """The backend should reject KV cache specs that cannot be gathered

315 - correctly by the ROCm kernel.

316 - """

314 + correctly by the ROCm kernel."""

315 from vllm.v1.attention.backends.rocm_aiter_fa import AiterFlashAttentionBackend

316 from vllm.v1.kv_cache_interface import FullAttentionSpec

317

@@ -333,8 +331,7 @@ def test_aiter_mha_backend_validates_kv_cache_block_size():

331

332 def test_aiter_mha_backend_supports_compute_capability_matches_mi3xx_probe():

333 """The backend should trust the ROCm MI3xx probe instead of the raw torch

336 - capability tuple.

337 - """

334 + capability tuple."""

335 from vllm.platforms.interface import DeviceCapability

336 from vllm.v1.attention.backends.rocm_aiter_fa import AiterFlashAttentionBackend

337

@@ -359,8 +356,7 @@ def test_aiter_mha_env_flags_control_enablement(

356 use_aiter, use_mha, expected, monkeypatch

357 ):

358 """Both the global AITER flag and the MHA-specific flag must be enabled

362 - before the ROCm MHA path is considered active.

363 - """

359 + before the ROCm MHA path is considered active."""

360 from vllm._aiter_ops import rocm_aiter_ops

361

362 _assert_aiter_supported()

@@ -378,8 +374,7 @@ def test_aiter_mha_env_flags_control_enablement(

374

375 def test_aiter_mha_platform_gate_matches_install_and_arch():

376 """The global AITER availability check should only open on ROCm MI3xx when

381 - the aiter package is installed.

382 - """

377 + the aiter package is installed."""

378 from vllm._aiter_ops import IS_AITER_FOUND, is_aiter_found_and_supported

379

380 assert is_aiter_found_and_supported() is (

@@ -650,8 +645,7 @@ def test_aiter_mha_multi_batch(num_heads, head_size, dtype):

645 @pytest.mark.parametrize("num_blocks", DIRECT_NUM_BLOCKS)

646 def test_aiter_fa_large_block_table_matches_reference(num_blocks):

647 """The direct paged-KV path should stay stable for both normal and very

653 - large block tables.

654 - """

648 + large block tables."""

649 atol = 2e-2

650 rtol = 2e-2

651 _assert_aiter_supported()

@@ -673,8 +667,7 @@ def test_aiter_fa_large_block_table_matches_reference(num_blocks):

667 @pytest.mark.skipif(not on_mi3xx(), reason="MI300/MI350 ROCm only")

668 def test_aiter_fa_sliding_window_matches_reference():

669 """The direct kernel should respect the same sliding-window causal mask as

676 - the naive reference implementation.

677 - """

670 + the naive reference implementation."""

671 atol = 2e-2

672 rtol = 2e-2

673 _assert_aiter_supported()

tests/kernels/attention/test_rocm_aiter_mla_sparse_metadata_sync.py

+2 -4

@@ -111,8 +111,7 @@ def _make_mixed_common_metadata():

111

112 def _patch_build_deps(monkeypatch, events=None):

113 """Stub the aiter kernel, triton helper and CUDA sync so ``build()`` runs

114 - on CPU.

115 - """

114 + on CPU."""

115

116 def fake_generate_sparse_seqlen_triton(

117 query_lens, seq_lens, cu_query_lens, topk_token, num_tokens, max_query_len

@@ -153,8 +152,7 @@ def test_build_populates_decode_only_split_fields(monkeypatch):

152

153 def test_build_populates_mixed_split_fields(monkeypatch):

154 """Mixed decode+prefill batch: split is reported, prefill fields stay

156 - default because this impl always runs the MQA path.

157 - """

155 + default because this impl always runs the MQA path."""

156 builder = _make_builder()

157 _patch_build_deps(monkeypatch)

158

tests/kernels/attention/test_triton_decode_attention.py

+2 -4

@@ -125,8 +125,7 @@ def _quantize_to_fp8(tensor: torch.Tensor):

125 @pytest.mark.parametrize("PAGE_SIZE", [1, 16])

126 def test_decode_attention_fp8(B, L, H_Q, H_KV, D_QK, D_V, CACHE_SIZE, PAGE_SIZE):

127 """Test FP8 KV cache path: quantize K/V to FP8, run kernel with scales,

128 - and compare against BF16 reference output.

129 - """

128 + and compare against BF16 reference output."""

129 assert CACHE_SIZE % PAGE_SIZE == 0

130 dtype = torch.bfloat16

131 seq_len = L

@@ -249,8 +248,7 @@ def test_decode_attention_cross_layer_view(H_Q, H_KV, D_QK, D_V, is_mla, PAGE_SI

248 """The kernel must honor the cache's page-dim stride, not assume pages are

249 packed back-to-back. A per-layer view into a cross-layer (block-major)

250 cache has stride(0) inflated by num_layers; outputs must match a

252 - contiguous cache holding the same data exactly.

253 - """

251 + contiguous cache holding the same data exactly."""

252 B = 3

253 seq_len = 1027

254 CACHE_SIZE = 16384

tests/kernels/core/test_fused_embed_norm.py

+2 -4

@@ -41,8 +41,7 @@ def _rmsnorm(x: torch.Tensor, w: torch.Tensor, eps: float) -> torch.Tensor:

41 def test_fused_embed_norm_matches_reference():

42 """Main-model fusion: the residual is the exact gather and the second output

43 is a correct RMSNorm. The norm matches a full-precision reference to ~2 bf16

44 - ulp (rtol 1e-2) -- that gap is bf16 rounding, not the kernel.

45 - """

44 + ulp (rtol 1e-2) -- that gap is bf16 rounding, not the kernel."""

45 set_random_seed(13)

46 table = torch.randn(VOCAB, HIDDEN, dtype=DTYPE, device="cuda")

47 ids = torch.randint(0, VOCAB, (NUM_TOKENS,), dtype=torch.int32, device="cuda")

@@ -61,8 +60,7 @@ def test_fused_embed_norm_matches_reference():

60 @torch.inference_mode()

61 def test_fused_embed_eh_norm_matches_reference():

62 """MTP fusion (folded gather) is bit-exact vs gathering the embeds and

64 - feeding the untouched model-local ``fused_eh_norm``.

65 - """

63 + feeding the untouched model-local ``fused_eh_norm``."""

64 set_random_seed(13)

65 table = torch.randn(VOCAB, HIDDEN, dtype=DTYPE, device="cuda")

66 ids = torch.randint(0, VOCAB, (NUM_TOKENS,), dtype=torch.int32, device="cuda")

tests/kernels/core/test_fused_q_kv_rmsnorm.py

+2 -4

@@ -62,8 +62,7 @@ def test_fused_q_kv_rmsnorm_outputs_are_packed(num_tokens: int):

62 a [1, q_size] slice of a [1, q_size + kv_size] buffer used to produce a

63 qr_out with row stride q_size + kv_size. Downstream dispatchers that

64 require packed row-major inputs then reject the tensor and silently fall

65 - back to a slower GEMM path on every decode step.

66 - """

65 + back to a slower GEMM path on every decode step."""

66 device = "cuda"

67 dtype = torch.bfloat16

68 q_size, kv_size = 192, 576

@@ -89,8 +88,7 @@ def test_fused_q_kv_rmsnorm_outputs_are_packed(num_tokens: int):

88 def test_fused_q_kv_rmsnorm_launches_past_grid_y_cap(num_tokens: int):

89 """Regression guard: grid used to be (2, num_tokens), hitting CUDA's

90 65535 grid-y cap at num_tokens >= 65536. The new grid (num_tokens, 2)

92 - lifts that bound to 2**31-1.

93 - """

91 + lifts that bound to 2**31-1."""

92 device = "cuda"

93 dtype = torch.bfloat16

94 q_size, kv_size = 192, 576

tests/kernels/core/test_fused_rms_norm_gated.py

+1 -2

@@ -2,8 +2,7 @@

2 # SPDX-FileCopyrightText: Copyright contributors to the vLLM project

3

4 """Tests that FusedRMSNormGated decomposes correctly under torch.compile,

5 -matching the eager triton kernel output.

6 -"""

5 +matching the eager triton kernel output."""

6

7 import pytest

8 import torch

tests/kernels/core/test_rocm_aiter_ops.py

+1 -2

@@ -579,8 +579,7 @@ def test_rocm_aiter_mla_dual_rmsnorm_group_quant_vs_sequential(m):

579

580 def test_rocm_aiter_mla_dual_rmsnorm_group_quant_strided_split_view():

581 """Split-view inputs, as produced by the DeepSeek-V4 q/kv split, match

582 - contiguous inputs.

583 - """

582 + contiguous inputs."""

583 require_aiter()

584 require_fp8()

585 from vllm._aiter_ops import rocm_aiter_ops

tests/kernels/core/test_vocab_parallel_embedding.py

+1 -2

@@ -87,8 +87,7 @@ def _shard(full_table, si, rows_per_rank, dtype):

87 @pytest.mark.parametrize("id_dtype", [torch.int32, torch.int64])

88 def test_matches_masked_gather(tp_size, vocab_size, org_vocab_size, hidden, id_dtype):

89 """Every rank's partial output is bit-exact with the eager path, including

90 - the added-vocab (LoRA) rows and the ids owned by another rank.

91 - """

90 + the added-vocab (LoRA) rows and the ids owned by another rank."""

91 set_random_seed(7)

92 dtype = torch.bfloat16

93 rows = pad_vocab_size(vocab_size, 64) // tp_size

tests/kernels/helion/test_autotune.py

+1 -2

@@ -59,8 +59,7 @@ class TestAutotuneDisabledKernel:

59

60 def test_autotune_disabled_kernel_produces_valid_config(self):

61 """Register a kernel with no configs (disabled), run autotune,

62 - verify it produces a valid helion.Config.

63 - """

62 + verify it produces a valid helion.Config."""

63 with dummy_kernel_registry(configs={}) as register:

64 wrapper = register(

65 "autotune_test_kernel",

tests/kernels/helion/test_register.py

+3 -6

@@ -598,8 +598,7 @@ class TestHelionKernelWrapper:

598 )

599 def test_init_eagerly_initializes_hop_path(self):

600 """Test that register_kernel eagerly builds the configured kernel

601 - on the HOP path (no custom op registration needed).

602 - """

601 + on the HOP path (no custom op registration needed)."""

602 from vllm.kernels.helion.utils import get_canonical_gpu_name

603

604 configs: dict[CaseKey, helion.Config] = {

@@ -634,8 +633,7 @@ class TestHelionKernelWrapper:

633 )

634 def test_init_eagerly_initializes(self):

635 """Test that register_kernel eagerly loads configs and detects GPU

637 - during construction so __call__ needs no further initialization.

638 - """

636 + during construction so __call__ needs no further initialization."""

637 from vllm.kernels.helion.utils import get_canonical_gpu_name

638

639 with (

@@ -963,8 +961,7 @@ class TestTorchCompileHOP:

961

962 def test_compiled_graph_contains_helion_hop(self):

963 """Verify torch.compile on a HelionKernelWrapper emits a

966 - helion_kernel_wrapper_mutation HOP node in the FX graph.

967 - """

964 + helion_kernel_wrapper_mutation HOP node in the FX graph."""

965 configs: dict[CaseKey, helion.Config] = {

966 CaseKey.default(): helion.Config(block_sizes=[4, 4])

967 }

tests/kernels/ir/test_layernorm.py

+1 -2

@@ -368,8 +368,7 @@ class TestFusedAddRMSNorm:

368 @pytest.mark.parametrize("provider", ["vllm_c"])

369 def test_inplace_semantics(self, dtype, n_tokens, hidden_size, epsilon, provider):

370 """Test that inplace implementations reuse inputs,

371 - for maybe_inplace overload but not for default overload.

372 - """

371 + for maybe_inplace overload but not for default overload."""

372 impl = ir.ops.fused_add_rms_norm.impls[provider]

373 if not impl.supported:

374 pytest.skip(f"{provider} impl not supported on this platform")

tests/kernels/mamba/cpu/test_cpu_gdn_ops.py

+3 -6

@@ -1055,8 +1055,7 @@ CONV_EQUIV_SEQ_LENS = [[1], [7], [64], [65], [1, 2, 3], [63, 64, 65], [128, 129]

1055

1056 def _conv_fp32_oracle(x, weight, bias, seq_lens, activation="silu"):

1057 """High-precision conv reference: everything in float32, no bf16 rounding.

1058 - x: [total_tokens, dim] (no initial state). Returns [total_tokens, dim].

1059 - """

1058 + x: [total_tokens, dim] (no initial state). Returns [total_tokens, dim]."""

1059 xf = x.float()

1060 wf = weight.float().unsqueeze(1)

1061 bf = bias.float()

@@ -1165,8 +1164,7 @@ def test_conv_cpp_no_worse_than_torch_vs_fp32(seq_lens):

1164 def test_conv_cpp_vnni_packed_matches_torch(seq_lens):

1165 """The exact runtime prefill sequence (VNNI-packed weight + SD-layout

1166 conv_state view + is_vnni=True) must match the torch fallback. Validates

1168 - the packing + layout handoff on any AVX-512BF16 CPU.

1169 - """

1167 + the packing + layout handoff on any AVX-512BF16 CPU."""

1168 x, weight, bias = _conv_inputs(sum(seq_lens))

1169 out_torch, _ = _run_prefill_torch(x, weight, bias, seq_lens)

1170 out_vnni, _ = _run_prefill_cpp(x, weight, bias, seq_lens, is_vnni=True)

@@ -1180,8 +1178,7 @@ def test_conv_cpp_vnni_packed_matches_torch(seq_lens):

1178 @torch.inference_mode()

1179 def test_conv_update_cpp_matches_torch(batch):

1180 """Decode conv: causal_conv1d_update_cpu matches causal_conv1d_update_torch,

1183 - including the in-place conv_state update (the next-step handoff).

1184 - """

1181 + including the in-place conv_state update (the next-step handoff)."""

1182 from vllm.model_executor.layers.mamba.ops.cpu.causal_conv1d import (

1183 causal_conv1d_update_torch,

1184 )

tests/kernels/mamba/test_mamba_ssm_configs.py

+3 -6

@@ -122,8 +122,7 @@ def test_fallback_when_no_config(monkeypatch, tmp_path):

122

123 def test_nearest_effective_batch_interpolation(monkeypatch, tmp_path):

124 """When effective_batch = batch*nheads is not an exact key, the closest

125 - key should be selected.

126 - """

125 + key should be selected."""

126 _write_config(

127 tmp_path,

128 dstate=32,

@@ -168,8 +167,7 @@ def test_nearest_effective_batch_interpolation(monkeypatch, tmp_path):

167

168 def test_non_dict_json_returns_none(monkeypatch, tmp_path):

169 """A valid JSON file that is not a dict (e.g. a list) must be ignored

171 - and return None rather than raising AttributeError.

172 - """

170 + and return None rather than raising AttributeError."""

171 device_name = get_ssm_device_name()

172 config_path = tmp_path / get_ssm_config_file_name(

173 _HEADDIM, 16, _CACHE_DTYPE, device_name

@@ -191,8 +189,7 @@ def test_non_dict_json_returns_none(monkeypatch, tmp_path):

189

190 def test_empty_config_falls_back_to_heuristic(monkeypatch, tmp_path):

191 """An empty JSON object {} must not crash min() — should fall back

194 - to the hard-coded heuristic.

195 - """

192 + to the hard-coded heuristic."""

193 _write_config(tmp_path, dstate=64, payload={})

194

195 monkeypatch.setenv("VLLM_TUNED_CONFIG_FOLDER", str(tmp_path))

tests/kernels/mamba/test_replayssm_prefill_decode_equivalence_mamba2.py

+1 -2

@@ -68,8 +68,7 @@ def _run_prefill_decode_equivalence(

68 the exact fp32 recurrence (and each other). All paths use the production dt

69 flow (raw dt + softplus + a per-head dt_bias), so this also checks prefill

70 and decode apply the softplus/bias preprocessing consistently. ``state_dtype``

71 - is the recurrent-state precision; ``act_dtype`` the activation/buffer one.

72 - """

71 + is the recurrent-state precision; ``act_dtype`` the activation/buffer one."""

72 device = current_platform.device_type

73 rtol, atol = _prefill_tolerances(act_dtype)

74 set_random_seed(seed)

tests/kernels/mamba/test_replayssm_standard_decode_mamba2.py

+2 -4

@@ -95,8 +95,7 @@ def _run_standard_decode(

95 ) -> None:

96 """Drive ``num_steps`` decode steps and check every path against the

97 trusted recurrence. ``state_dtype`` is the recurrent-state precision;

98 - ``act_dtype`` is the activation/buffer precision.

99 - """

98 + ``act_dtype`` is the activation/buffer precision."""

99 device = "cuda"

100 both_fp32 = state_dtype == torch.float32 and act_dtype == torch.float32

101 rtol, atol = _tolerances(torch.float32 if both_fp32 else torch.bfloat16)

@@ -540,8 +539,7 @@ def test_replayssm_standard_decode_tp_head_shard_equivalence(

539 separable: one run over all heads must equal concatenating ``tp`` independent

540 per-rank runs on ``nheads // tp`` heads and ``ngroups // tp`` groups. This

541 guards the per-rank divisors and group->head mapping the state-shape wiring

543 - relies on. The real TP1==TP2 engine check lives in the v1/e2e suite.

544 - """

542 + relies on. The real TP1==TP2 engine check lives in the v1/e2e suite."""

543 state_dtype, act_dtype = precision

544 nheads, headdim, dstate, ngroups = geometry

545 assert nheads % tp == 0 and ngroups % tp == 0

tests/kernels/moe/modular_kernel_tools/common.py

+1 -2

@@ -167,8 +167,7 @@ class Config:

167

168 def fe_supports_quant_scheme(self) -> bool:

169 """Check if the fused experts class supports this quant config.

170 - See https://github.com/ROCm/aiter/issues/2419 for AITER gaps.

171 - """

170 + See https://github.com/ROCm/aiter/issues/2419 for AITER gaps."""

171 if self.quant_config is None or self.quant_dtype is None:

172 return True

173 if self.quant_dtype != torch.float8_e4m3fn:

tests/kernels/moe/test_b12x.py

+1 -2

@@ -468,8 +468,7 @@ def test_deepseek_v4_flashinfer_cutlass_falls_through_to_w4a8(

468 monkeypatch: pytest.MonkeyPatch,

469 ) -> None:

470 """Explicit flashinfer_cutlass must try the W4A8 variant when the BF16

471 - variant is unsupported (the BF16 variant is gated to SM90).

472 - """

471 + variant is unsupported (the BF16 variant is gated to SM90)."""

472 from vllm.model_executor.layers.fused_moe.experts.flashinfer_cutlass_moe import (

473 FlashInferExperts,

474 )

tests/kernels/moe/test_batched_moe.py

+3 -6

@@ -129,8 +129,7 @@ def test_batched_mm(

129 per_act_token_quant: bool,

130 ):

131 """Note: float8_e4m3fn is not supported on CUDA architecture < 89,

132 - and those tests will be skipped on unsupported hardware.

133 - """

132 + and those tests will be skipped on unsupported hardware."""

133 set_random_seed(7)

134

135 use_fp8_w8a8 = dtype == torch.float8_e4m3fn

@@ -269,8 +268,7 @@ def test_fused_moe_batched_experts(

268 workspace_init,

269 ):

270 """Note: float8_e4m3fn is not supported on CUDA architecture < 89,

272 - and those tests will be skipped on unsupported hardware.

273 - """

271 + and those tests will be skipped on unsupported hardware."""

272 set_random_seed(7)

273

274 use_fp8_w8a8 = dtype == torch.float8_e4m3fn

@@ -651,8 +649,7 @@ def test_batched_experts_masked_activation_end_to_end(

649 def test_batched_experts_end_to_end(m, n, k, e, topk):

650 """End-to-end BatchedTritonExperts via the reference (no-comms)

651 BatchedPrepareAndFinalize, validated against a torch reference. Exercises

654 - the device-enablement path.

655 - """

652 + the device-enablement path."""

653 from vllm.v1.worker.workspace import init_workspace_manager

654

655 set_random_seed(7)

tests/kernels/moe/test_block_int8.py

+2 -4

@@ -54,8 +54,7 @@ SEEDS = [0]

54 # For test

55 def torch_w8a8_block_int8_moe(a, w1, w2, w1_s, w2_s, score, topk, block_shape):

56 """This function performs fused moe with block-wise quantization using

57 - native torch.

58 - """

57 + native torch."""

58 B, D = a.shape

59 a = a.view(B, -1, D).repeat(1, topk, 1).reshape(-1, D)

60 out = torch.zeros(B * topk, w2.shape[1], dtype=a.dtype, device=a.device)

@@ -98,8 +97,7 @@ def setup_cuda():

97 @torch.inference_mode()

98 def test_w8a8_block_int8_fused_moe(M, N, K, E, topk, block_size, dtype, seed):

99 """Tests the fused_moe kernel with W8A8 INT8 block quantization against a

101 - native torch reference.

102 - """

100 + native torch reference."""

101 torch.manual_seed(seed)

102

103 a = torch.randn((M, K), dtype=dtype) / 10

tests/kernels/moe/test_cpu_fused_moe.py

+2 -4

@@ -397,8 +397,7 @@ UNALIGNED_INTERMEDIATE_DIM = 176

397 class _StubMoELayer(torch.nn.Module):

398 """Minimal stand-in for the real MoE layer module, exposing just what

399 the unquantized CPU experts read/replace (w13_weight, w2_weight, the

400 - router configuration, and optionally w13_bias/w2_bias).

401 - """

400 + router configuration, and optionally w13_bias/w2_bias)."""

401

402 def __init__(

403 self,

@@ -501,8 +500,7 @@ def test_cpu_fused_moe_unaligned_intermediate_size(

500 act: MoEActivation,

501 ):

502 """CPU kernels handle unaligned intermediate sizes by zero-padding the

504 - weights before prepacking.

505 - """

503 + weights before prepacking."""

504 set_random_seed(0)

505 batch_size = 64

506 intermediate_size = UNALIGNED_INTERMEDIATE_DIM

tests/kernels/moe/test_flashinfer.py

+1 -2

@@ -464,8 +464,7 @@ def test_trtllm_fp8_swiglu_clamp_support(

464 ):

465 """FlashInfer >= 0.6.18 applies the SwiGLU clamp for both block-scaled

466 kernels with a SwiGLU activation (DeepSeek-V4 sets swiglu_limit); the

467 - per-tensor kernel has no clamp, and Relu2 rejects the parameters.

468 - """

467 + per-tensor kernel has no clamp, and Relu2 rejects the parameters."""

468

469 class _Experts(TrtLlmFp8ExpertsModular):

470 @staticmethod

tests/kernels/moe/test_gpt_oss_triton_kernels.py

+1 -2

@@ -377,8 +377,7 @@ def test_unit_shuffle():

377 @pytest.mark.parametrize("n_experts,topk", [(32, 4), (128, 4)])

378 def test_routing_data_from_sparse_topk_parity(n_tokens, n_experts, topk):

379 """routing_data_from_sparse_topk must produce routing structures

380 - identical to make_routing_data for the same topk result.

381 - """

380 + identical to make_routing_data for the same topk result."""

381 from vllm.model_executor.layers.fused_moe.experts import (

382 gpt_oss_triton_kernels_moe as gptoss_moe,

383 )

tests/kernels/moe/test_marlin_vs_trtllm_mxint4.py

+1 -2

@@ -289,8 +289,7 @@ def test_marlin_vs_trtllm_mxint4_moe_kimik2(monkeypatch, m, n, k, e, topk, group

289 @torch.inference_mode()

290 def test_flashinfer_trtllm_mxint4_moe_wrapper(m, n, k, e, topk):

291 """Test that the flashinfer_trtllm_mxint4_moe wrapper matches the raw

292 - trtllm_mxint4_block_scale_moe kernel call.

293 - """

292 + trtllm_mxint4_block_scale_moe kernel call."""

293 pytest.importorskip("flashinfer")

294 from flashinfer import RoutingMethodType

295 from flashinfer.fused_moe import trtllm_mxint4_block_scale_moe

tests/kernels/moe/test_modular_kernel_combinations.py

+1 -2

@@ -352,8 +352,7 @@ def test_modular_kernel_combinations_singlegpu(

352 workspace_init,

353 ):

354 """Note: float8_e4m3fn is not supported on CUDA architecture < 89,

355 - and those tests will be skipped on unsupported hardware.

356 - """

355 + and those tests will be skipped on unsupported hardware."""

356 config = Config(

357 Ms=Ms,

358 K=k,

tests/kernels/moe/test_moe_kernel_oracle.py

+1 -2

@@ -21,8 +21,7 @@ from vllm.model_executor.layers.fused_moe.oracle.unquantized import (

21

22 class TestUnquantizedDelegation:

23 """UnquantizedMoEKernelOracle methods must delegate to the existing

24 - module-level functions; behaviour is bit-identical.

25 - """

24 + module-level functions; behaviour is bit-identical."""

25

26 def test_make_kernel_delegates(self) -> None:

27 quant_config = object()

tests/kernels/moe/test_moe_layer.py

+1 -2

@@ -315,8 +315,7 @@ def tp_chunk_gate_up(

315 device: torch.device | int | None = None,

316 ) -> torch.Tensor:

317 """TP-chunk a combined [gate; up] weight, splitting each half separately

318 - so every rank gets a portion of both gate and up.

319 - """

318 + so every rank gets a portion of both gate and up."""

319 half = w.shape[dim] // 2

320 gate = chunk_by_rank(

321 w.narrow(dim, 0, half), tp_rank, tp_size, dim=dim, device=device

tests/kernels/moe/test_mxfp8_aiter_backend_selection.py

+2 -4

@@ -130,8 +130,7 @@ def test_is_supported_config(present, ep_size, supported, reason_substr):

130

131 def test_explicit_moe_backend_aiter():

132 """--moe-backend aiter: returns FlyDSL when usable (TP or EP), else a clear

133 - ValueError when the flydsl package is missing.

134 - """

133 + ValueError when the flydsl package is missing."""

134 with _gfx950(), _flydsl_installed(True):

135 assert (

136 _select_kernel_cls(Fp8MoeBackend.AITER_MXFP8, _config(1))

@@ -163,8 +162,7 @@ def test_gfx950_picks_aiter():

162

163 def test_gfx942_picks_emulation():

164 """Flydsl unusable (e.g. gfx942, no FlyDSL support) -> native Triton

166 - dot_scaled backend wins instead.

167 - """

165 + dot_scaled backend wins instead."""

166 with patch(f"{_AITER_MOD}.current_platform.supports_mx", return_value=False):

167 backend, experts_cls = select_mxfp8_moe_backend(_config())

168 assert backend is Fp8MoeBackend.EMULATION

tests/kernels/moe/test_rocm_aiter_moe.py

+11 -22

@@ -249,8 +249,7 @@ def _make_topk_ids(

249 device: str = "cuda",

250 ) -> torch.Tensor:

251 """Generate distinct expert IDs per token with the same top-k shape as

252 - production routing.

253 - """

252 + production routing."""

253 router_logits = torch.randn(num_tokens, num_experts, device=device)

254 _, topk_ids = torch.topk(torch.softmax(router_logits, dim=-1), k=topk, dim=-1)

255 return topk_ids.to(torch.int32)

@@ -444,8 +443,7 @@ def test_aiter_fused_moe_custom_op_registered():

443

444 def test_aiter_asm_moe_tkw1_custom_op_registered():

445 """The tkw1 custom op should stay registered for FP8 apply-router-weight

447 - paths.

448 - """

446 + paths."""

447 _assert_aiter_supported()

448 import vllm._aiter_ops as aiter_ops # noqa: F401

449

@@ -455,8 +453,7 @@ def test_aiter_asm_moe_tkw1_custom_op_registered():

453

454 def test_aiter_fused_moe_fake_tensor_support():

455 """The fused-MoE op should preserve fake-tensor compatibility for

458 - torch.compile-style tracing.

459 - """

456 + torch.compile-style tracing."""

457 _assert_aiter_supported()

458 import vllm._aiter_ops # noqa: F401

459

@@ -519,8 +516,7 @@ def test_aiter_moe_enablement_follows_env(

516 monkeypatch: pytest.MonkeyPatch,

517 ):

518 """The fused-MoE gate should depend only on the main AITER toggle and the

522 - MoE-specific toggle.

523 - """

519 + MoE-specific toggle."""

520 from vllm._aiter_ops import rocm_aiter_ops

521

522 _assert_aiter_supported()

@@ -551,8 +547,7 @@ def test_aiter_moe_shared_experts_enablement_follows_env(

547 monkeypatch: pytest.MonkeyPatch,

548 ):

549 """Shared-expert fusion should only be enabled when the fused-MoE path is

554 - enabled too.

555 - """

550 + enabled too."""

551 from vllm._aiter_ops import rocm_aiter_ops

552

553 _assert_aiter_supported()

@@ -693,8 +688,7 @@ def test_aiter_mxfp4_quant_scheme_support_matches_gfx950():

688 @pytest.mark.skipif(not on_gfx950(), reason="gfx950 ROCm only")

689 def test_aiter_fused_moe_mi350_mxfp4_w4a16_accuracy():

690 """The gfx950 AITER MXFP4 W4A16 MoE path should match the dequantized

696 - MXFP4 reference.

697 - """

691 + MXFP4 reference."""

692 from vllm.model_executor.layers.fused_moe.activation import MoEActivation

693 from vllm.model_executor.layers.fused_moe.experts.rocm_aiter_moe import (

694 rocm_aiter_fused_experts,

@@ -743,8 +737,7 @@ def test_aiter_fused_moe_mi350_mxfp4_w4a16_accuracy():

737 @pytest.mark.skipif(not on_gfx950(), reason="gfx950 ROCm only")

738 def test_aiter_fused_moe_mi350_mxfp4_w4a16_determinism():

739 """The gfx950 AITER MXFP4 W4A16 MoE path should stay bitwise

746 - deterministic.

747 - """

740 + deterministic."""

741 from vllm.model_executor.layers.fused_moe.activation import MoEActivation

742 from vllm.model_executor.layers.fused_moe.experts.rocm_aiter_moe import (

743 rocm_aiter_fused_experts,

@@ -837,8 +830,7 @@ def test_aiter_fused_moe_bf16_accuracy(

830 intermediate_dim: int,

831 ):

832 """The ROCm AITER fused-MoE BF16 path should match the float32 reference

840 - on representative shapes.

841 - """

833 + on representative shapes."""

834 from vllm.model_executor.layers.fused_moe.experts.rocm_aiter_moe import (

835 ActivationMethod,

836 QuantMethod,

@@ -883,8 +875,7 @@ def test_aiter_fused_moe_bf16_accuracy(

875

876 def test_aiter_fused_moe_gelu_accuracy():

877 """The GELU activation variant should stay aligned with the float32

886 - reference.

887 - """

878 + reference."""

879 from vllm.model_executor.layers.fused_moe.experts.rocm_aiter_moe import (

880 ActivationMethod,

881 QuantMethod,

@@ -929,8 +920,7 @@ def test_aiter_fused_moe_gelu_accuracy():

920

921 def test_aiter_fused_moe_determinism():

922 """The BF16 fused-MoE kernel should stay bitwise deterministic for the

932 - same inputs.

933 - """

923 + same inputs."""

924 from vllm.model_executor.layers.fused_moe.experts.rocm_aiter_moe import (

925 ActivationMethod,

926 QuantMethod,

@@ -1085,8 +1075,7 @@ def test_aiter_fused_moe_mi3xx_bf16_accuracy():

1075 )

1076 def test_aiter_fused_moe_mi3xx_fp8_accuracy():

1077 """The MI3xx FP8 per-tensor MoE path should stay within the measured FP8

1088 - error budget.

1089 - """

1078 + error budget."""

1079 from tests.kernels.moe.utils import make_test_weights

1080 from vllm._aiter_ops import rocm_aiter_ops

1081 from vllm.model_executor.layers.fused_moe.experts.rocm_aiter_moe import (

tests/kernels/moe/test_routed_experts_capture_monolithic.py

+8 -16

@@ -185,8 +185,7 @@ def test_trtllm_bf16_monolithic_routing_replay_records_valid_experts(

185 ) -> None:

186 """The capture callback should receive the int16 routed-expert IDs the

187 kernel actually used, the values should be valid expert indices, and

188 - each token should pick ``top_k`` distinct experts.

189 - """

188 + each token should pick ``top_k`` distinct experts."""

189 if top_k > _DSV3_N_GROUP * _DSV3_TOPK_GROUP:

190 pytest.skip(

191 f"DSV3 requires top_k <= n_group * topk_group "

@@ -265,8 +264,7 @@ def test_trtllm_bf16_monolithic_routing_replay_non_dsv3(

264 routing_method: RoutingMethodType,

265 ) -> None:

266 """Routing replay works for non-DeepSeekV3 routing methods too.

268 - FlashInfer's ``routing_replay_out`` is routing-method-agnostic.

269 - """

267 + FlashInfer's ``routing_replay_out`` is routing-method-agnostic."""

268 torch.manual_seed(0)

269 device = torch.device("cuda:0")

270

@@ -319,8 +317,7 @@ def test_trtllm_bf16_monolithic_routing_replay_non_dsv3(

317

318 def test_trtllm_bf16_monolithic_capture_disabled_skips_buffer_alloc() -> None:

319 """With no callback installed the kernel should not see a

322 - ``routing_replay_out`` tensor — verify the helper short-circuits.

323 - """

320 + ``routing_replay_out`` tensor — verify the helper short-circuits."""

321 torch.manual_seed(0)

322 device = torch.device("cuda:0")

323 experts, _, _ = _make_bf16_monolithic_experts(

@@ -342,8 +339,7 @@ def test_trtllm_bf16_monolithic_capture_disabled_skips_buffer_alloc() -> None:

339 def test_trtllm_bf16_monolithic_supports_capture_for_all_routing() -> None:

340 """FlashInfer's ``routing_replay_out`` is supported by all routing

341 methods, so ``supports_routing_replay_capture`` should be True

345 - regardless of routing method.

346 - """

342 + regardless of routing method."""

343 device = torch.device("cuda:0")

344 for routing_method in (

345 RoutingMethodType.DeepSeekV3,

@@ -365,8 +361,7 @@ def test_trtllm_bf16_monolithic_supports_capture_for_all_routing() -> None:

361

362 def test_trtllm_bf16_monolithic_capture_buffer_shape_and_dtype() -> None:

363 """When capture is installed, the allocated buffer is int16 and shaped

368 - ``(num_tokens, experts_per_token)``.

369 - """

364 + ``(num_tokens, experts_per_token)``."""

365 device = torch.device("cuda:0")

366 experts, _, _ = _make_bf16_monolithic_experts(

367 num_experts=_DSV3_NUM_EXPERTS,

@@ -597,8 +592,7 @@ def test_trtllm_fp8_block_scale_monolithic_routing_replay_records_valid_experts(

592 top_k: int,

593 ) -> None:

594 """End-to-end: ``TrtLlmFp8ExpertsMonolithic`` (DeepSeekFp8 block-scale

600 - path, DSV3 routing) captures valid expert IDs.

601 - """

595 + path, DSV3 routing) captures valid expert IDs."""

596 if top_k > _DSV3_N_GROUP * _DSV3_TOPK_GROUP:

597 pytest.skip(

598 f"DSV3 requires top_k <= n_group * topk_group "

@@ -784,8 +778,7 @@ def _run_nvfp4_monolithic(

778 routing_bias: torch.Tensor,

779 ) -> torch.Tensor:

780 """The monolithic NVFP4 apply expects packed fp4 hidden states + the

787 - matching fp8 per-block scale stored in the ``a1q_scale`` slot.

788 - """

781 + matching fp8 per-block scale stored in the ``a1q_scale`` slot."""

782 # Stash the weight tensors on the experts in the locations the apply()

783 # implementation reads from (it pulls them from quant_config / scales

784 # already; w1/w2 come in as args).

@@ -813,8 +806,7 @@ def test_trtllm_nvfp4_monolithic_routing_replay_records_valid_experts(

806 top_k: int,

807 ) -> None:

808 """End-to-end: ``TrtLlmNvFp4ExpertsMonolithic`` captures valid expert IDs

816 - on the DSV3 routing path.

817 - """

809 + on the DSV3 routing path."""

810 if top_k > _DSV3_N_GROUP * _DSV3_TOPK_GROUP:

811 pytest.skip(

812 f"DSV3 requires top_k <= n_group * topk_group "

tests/kernels/moe/test_routing.py

+1 -2

@@ -250,8 +250,7 @@ def assert_aiter_routing_valid(

250 the Python baseline (different group selection, scoring internals),

251 so numerical comparison is not meaningful. Instead we verify the

252 outputs satisfy the routing contract: correct shapes, valid expert

253 - IDs, non-negative weights, and proper normalization.

254 - """

253 + IDs, non-negative weights, and proper normalization."""

254 n_tokens = topk_weights.shape[0]

255

256 # Shape

tests/kernels/moe/test_routing_simulator.py

+2 -4

@@ -76,8 +76,7 @@ def test_basic_functionality(

76

77 def test_routing_strategy_integration(monkeypatch, device):

78 """Test that the routing strategy environment variable works with

79 - FusedMoEFactory.

80 - """

79 + FusedMoEFactory."""

80 pytest.importorskip("vllm.model_executor.layers.fused_moe.layer")

81

82 import vllm.envs as envs

@@ -158,8 +157,7 @@ def test_routing_strategy_integration(monkeypatch, device):

157

158 def test_distribution_based_routing_with_custom_strategy():

159 """Test registering and using DistributionBasedRouting with custom

161 - parameters.

162 - """

160 + parameters."""

161 device = torch.device("cuda" if torch.cuda.is_available() else "cpu")

162

163 # Register custom distribution-based strategy

tests/kernels/moe/test_silu_mul_per_token_group_quant_fp8_colmajor.py

+1 -2

@@ -70,8 +70,7 @@ def reference_with_clamp(

70 ) -> tuple[torch.Tensor, torch.Tensor]:

71 """Pre-clamp inputs (gate from above, up symmetric) at the input dtype to

72 match the C++ compute() template, then run the standard silu_and_mul +

73 - quant reference.

74 - """

73 + quant reference."""

74 N_2 = x.size(1) // 2

75 dtype = x.dtype

76 gate = x[..., :N_2].to(torch.float32).clamp(max=clamp_limit).to(dtype)

tests/kernels/moe/test_triton_moe_ptpc_fp8.py

+2 -4

@@ -28,8 +28,7 @@ if current_platform.is_fp8_fnuz():

28

29 def native_w8a8_per_token_matmul(A, B, As, Bs, output_dtype=torch.float16):

30 """Matrix multiplication function that supports per-token input

31 - quantization and per-column weight quantization

32 - """

31 + quantization and per-column weight quantization"""

32 A = A.to(torch.float32)

33 B = B.to(torch.float32)

34

@@ -57,8 +56,7 @@ def fp8_mask(a, mask):

56

57 def torch_w8a8_per_column_moe(a, w1, w2, w1_s, w2_s, score, topk):

58 """This function performs fused moe with per-column int8

60 - quantization using native torch.

61 - """

59 + quantization using native torch."""

60 B, D = a.shape

61 # Perform per-token quantization

62 a_q, a_s = ops.scaled_fp8_quant(a, use_per_token_if_dynamic=True)

tests/kernels/moe/test_unquantized_backend_selection.py

+2 -4

@@ -150,8 +150,7 @@ def test_select_default_backend_by_platform(

150 expected_backend,

151 ):

152 """Test default backend selection per platform with all optional

153 - accelerators (FlashInfer, AITER) disabled.

154 - """

153 + accelerators (FlashInfer, AITER) disabled."""

154 with patch(

155 "vllm.model_executor.layers.fused_moe.oracle.unquantized.current_platform"

156 ) as mock_platform:

@@ -445,8 +444,7 @@ def test_select_cuda_flashinfer_cutlass_backend(

444 mock_has_flashinfer,

445 ):

446 """Test CUDA backend selection when FlashInfer TRTLLM is not available

448 - and FlashInfer CUTLASS is available.

449 - """

447 + and FlashInfer CUTLASS is available."""

448 with (

449 patch.object(current_platform, "is_cuda", return_value=True),

450 patch.object(current_platform, "is_rocm", return_value=False),

tests/kernels/quant_utils.py

+1 -2

@@ -159,8 +159,7 @@ def native_per_token_group_quant_fp8(

159 x, group_size, eps=1e-10, dtype=torch.float8_e4m3fn

160 ):

161 """Function to perform per-token-group quantization on an input tensor

162 - `x` using native torch.

163 - """

162 + `x` using native torch."""

163 assert x.shape[-1] % group_size == 0, (

164 "the last dimension of `x` must be divisible by `group_size`"

165 )

tests/kernels/quantization/test_aiter_hipb_mm_linear_kernel.py

+1 -2

@@ -304,8 +304,7 @@ def test_hipb_mm_kernel_forward_matches_raw_aiter_hipb_mm(enable_hipb_mm_kernel)

304

305 def test_hipb_mm_kernel_forward_accuracy(enable_hipb_mm_kernel):

306 """Kernel output should match a dequantized fp32 reference within

307 - fp8 per-token / per-channel quantization noise.

308 - """

307 + fp8 per-token / per-channel quantization noise."""

308 weight_shape = (512, 4096) # (N, K)

309 num_tokens = 32

310 _check_bpreshuffle_runtime_support(weight_shape, num_tokens=num_tokens)

tests/kernels/quantization/test_int8_kernel.py

+2 -4

@@ -30,8 +30,7 @@ if current_platform.is_cuda_alike() and not current_platform.has_device_capabili

30

31 def native_w8a8_per_token_matmul(A, B, As, Bs, output_dtype=torch.float16):

32 """Matrix multiplication function that supports per-token input

33 - quantization and per-column weight quantization

34 - """

33 + quantization and per-column weight quantization"""

34 A = A.to(torch.float32)

35 B = B.to(torch.float32)

36

@@ -54,8 +53,7 @@ def native_w8a8_per_token_matmul(A, B, As, Bs, output_dtype=torch.float16):

53

54 def torch_w8a8_per_column_moe(a, w1, w2, w1_s, w2_s, topk, topk_weight, topk_ids):

55 """This function performs fused moe with per-column int8 quantization

57 - using native torch.

58 - """

56 + using native torch."""

57 B, D = a.shape

58 # Perform per-token quantization

59 a_q, a_s = per_token_quant_int8(a)

tests/kernels/quantization/test_marlin_tile_padding.py

+5 -10

@@ -360,8 +360,7 @@ def test_gptq_marlin_padded_round_trip(shape, group_size):

360 def test_fp8_block_marlin_padded_round_trip(shape):

361 """Block-quantized FP8 (e.g. Nemotron NVFP4 checkpoints' FP8 layers):

362 group_size=128 exercises the lcm K-alignment in marlin_padded_nk and the

363 - weight_scale_inv group-wise scale padding.

364 - """

363 + weight_scale_inv group-wise scale padding."""

364 size_n, size_k = shape

365 block = 128

366 dtype = torch.float16

@@ -411,8 +410,7 @@ def test_fp8_block_marlin_padded_round_trip(shape):

410 @pytest.mark.parametrize("shape", [(200, 288), (4640, 512)])

411 def test_mxfp8_marlin_padded_round_trip(shape):

412 """MXFP8 exercises the e8m0 scale path, where padded 0.0 scales clamp to

414 - 2^-127 instead of zero and must still contribute nothing.

415 - """

413 + 2^-127 instead of zero and must still contribute nothing."""

414 size_n, size_k = shape

415 group_size = 32

416 # The e8m0-scale Marlin kernels are only instantiated for bf16 activations.

@@ -459,8 +457,7 @@ def test_mxfp8_marlin_padded_round_trip(shape):

457 @pytest.mark.parametrize("shape", [(200, 512), (4640, 512)])

458 def test_awq_zp_marlin_padded_round_trip(shape):

459 """AWQ-style uint4 with runtime zero-points, padded the way

462 - MarlinLinearKernel does: padded columns rely on (q=0 - zp=0) * scale=0.

463 - """

460 + MarlinLinearKernel does: padded columns rely on (q=0 - zp=0) * scale=0."""

461 size_n, size_k = shape

462 group_size = 128

463 dtype = torch.float16

@@ -695,8 +692,7 @@ def test_check_moe_marlin_supports_layer_padding():

692 @pytest.mark.parametrize("shape", [(96, 256, 8), (160, 512, 4)])

693 def test_fp8_marlin_moe_padded_round_trip(shape, quant):

694 """FP8 weight-only MoE: pad a tile-misaligned intermediate and check the

698 - real prepare + fused_marlin_moe against the dequantized reference.

699 - """

695 + real prepare + fused_marlin_moe against the dequantized reference."""

696 from tests.kernels.utils import torch_experts

697 from vllm.config import VllmConfig, set_current_vllm_config

698 from vllm.model_executor.layers.fused_moe import fused_topk

@@ -781,8 +777,7 @@ def test_fp8_marlin_moe_padded_round_trip(shape, quant):

777 @pytest.mark.parametrize("shape", [(96, 256, 8), (160, 512, 4)])

778 def test_mxfp8_marlin_moe_padded_round_trip(shape):

779 """MXFP8 weight-only MoE round-trip at a tile-misaligned intermediate, with

784 - unit e8m0 scales so the reference is the exact fp8 dequant.

785 - """

780 + unit e8m0 scales so the reference is the exact fp8 dequant."""

781 from tests.kernels.utils import torch_experts

782 from vllm.config import VllmConfig, set_current_vllm_config

783 from vllm.model_executor.layers.fused_moe import fused_topk

tests/kernels/quantization/test_mxfp4_kernel_selection.py

+6 -12

@@ -72,8 +72,7 @@ def test_true_w4a4_kernels_reject_unset_activation(kernel_cls):

72 @pytest.mark.parametrize("kernel_cls", _TRUE_W4A4_KERNELS)

73 def test_true_w4a4_kernels_reject_explicit_non_mxfp4_activation(kernel_cls):

74 """FlashInfer/XPU/Aiter quantize activations to MXFP4 internally, so an

75 - explicit request for a different activation format must be rejected.

76 - """

75 + explicit request for a different activation format must be rejected."""

76 config = MxFp4LinearLayerConfig(activation_quant_key=kMxfp6E3M2Dynamic)

77 can_implement, reason = kernel_cls.can_implement(config)

78 assert not can_implement

@@ -88,8 +87,7 @@ def test_weight_only_kernels_accept_unquantized_or_mxfp4_activation(

87 """Marlin/Humming never quantize activations, so an unset activation key,

88 or one that already describes MXFP4-shaped data, is tolerated. When an

89 activation key is explicitly set, a warning must be logged noting that it

91 - is ignored, since these kernels are weight-only (A16).

92 - """

90 + is ignored, since these kernels are weight-only (A16)."""

91 config = MxFp4LinearLayerConfig(activation_quant_key=activation_quant_key)

92 with patch(f"{kernel_cls.__module__}.logger.warning_once") as warning_once:

93 can_implement, reason = kernel_cls.can_implement(config)

@@ -118,8 +116,7 @@ def test_weight_only_kernels_reject_non_mxfp4_activation(kernel_cls):

116 )

117 def test_emulation_kernel_accepts_any_config(activation_quant_key, monkeypatch):

118 """EmulationMxfp4LinearKernel is the universal fallback: it must accept

121 - every supported activation format.

122 - """

119 + every supported activation format."""

120 # `EmulationMxfp4LinearKernel.can_implement` gates on `has_quark()`,

121 # which we are not testing here.

122 monkeypatch.setattr(

@@ -137,8 +134,7 @@ def test_emulation_kernel_accepts_any_config(activation_quant_key, monkeypatch):

134

135 def test_emulation_kernel_derives_quant_dequant_func_from_config(monkeypatch):

136 """quant_dequant_func must be derived purely from the config's activation

140 - QuantKey, not set externally.

141 - """

137 + QuantKey, not set externally."""

138 # `EmulationMxfp4LinearKernel.can_implement` gates on `has_quark()`,

139 # which we are not testing here.

140 monkeypatch.setattr(

@@ -162,8 +158,7 @@ def test_emulation_kernel_derives_quant_dequant_func_from_config(monkeypatch):

158

159 def test_aiter_kernel_is_supported_requires_native_mx_support():

160 """AiterMxfp4LinearKernel must not be selected on platforms without

165 - native MX compute, even if AITER itself is importable.

166 - """

161 + native MX compute, even if AITER itself is importable."""

162 with patch(

163 "vllm.model_executor.kernels.linear.mxfp4.aiter.current_platform.supports_mx",

164 return_value=False,

@@ -199,8 +194,7 @@ class OOTMxFp4LinearKernel(MxFp4LinearKernel):

194 @patch("vllm.model_executor.kernels.linear.current_platform")

195 def test_init_mxfp4_linear_kernel_dispatches_to_registered_kernel(platform_mock):

196 """init_mxfp4_linear_kernel should select a registered kernel that

202 - reports itself as supported, and construct it with a fresh config.

203 - """

197 + reports itself as supported, and construct it with a fresh config."""

198 platform_mock._enum = PlatformEnum.OOT

199 register_linear_kernel(OOTMxFp4LinearKernel, PlatformEnum.OOT, "mxfp4")

200

tests/kernels/quantization/test_mxfp4_triton_ep.py

+1 -2

@@ -16,8 +16,7 @@ import torch

16

17 class TestTritonMoeForwardExpertMap:

18 """Test that triton_kernel_moe_forward applies expert_map remapping

19 - when expert_map is provided (EP active).

20 - """

19 + when expert_map is provided (EP active)."""

20

21 def test_expert_map_remap(self):

22 device = "cuda" if torch.cuda.is_available() else "cpu"

tests/kernels/quantization/test_mxfp6_kernel_selection.py

+3 -6

@@ -41,8 +41,7 @@ def test_can_implement_is_abstract():

41

42 def test_emulation_kernel_rejects_non_mxfp6_weights():

43 """EmulationMxfp6LinearKernel must not implement a non-MXFP6 weight

44 - format.

45 - """

44 + format."""

45 config = MxFp6LinearLayerConfig(weight_quant_key=kMxfp4Static)

46 can_implement, reason = EmulationMxfp6LinearKernel.can_implement(config)

47 assert not can_implement

@@ -58,8 +57,7 @@ def test_emulation_kernel_accepts_any_supported_config(

57 weight_quant_key, activation_quant_key

58 ):

59 """EmulationMxfp6LinearKernel is the only backend today: it must accept

61 - every supported weight/activation format combination.

62 - """

60 + every supported weight/activation format combination."""

61 config = MxFp6LinearLayerConfig(

62 weight_quant_key=weight_quant_key, activation_quant_key=activation_quant_key

63 )

@@ -104,8 +102,7 @@ class OOTMxFp6LinearKernel(MxFp6LinearKernel):

102 def test_init_mxfp6_linear_kernel_dispatches_to_registered_kernel(platform_mock):

103 """init_mxfp6_linear_kernel should select a registered kernel that

104 reports itself as supported/able to implement the given config, and

107 - construct it with that exact config.

108 - """

105 + construct it with that exact config."""

106 platform_mock._enum = PlatformEnum.OOT

107 register_linear_kernel(OOTMxFp6LinearKernel, PlatformEnum.OOT, "mxfp6")

108

tests/kernels/quantization/test_per_token_group_quant.py

+3 -6

@@ -118,8 +118,7 @@ def test_per_token_group_quant_fp8_packed(

118 num_tokens, hidden_dim, group_size, poisoned_scales

119 ):

120 """Test the packed DeepGEMM quantization kernel against the Triton

121 - reference (row-major, UE8M0 scales).

122 - """

121 + reference (row-major, UE8M0 scales)."""

122 device = "cuda"

123 torch.manual_seed(42)

124

@@ -257,8 +256,7 @@ def test_per_token_group_quant_fp8_packed_all_zero():

256 def test_per_token_group_quant_fp8_packed_mantissa_rounds_up():

257 """Inputs whose absmax/max_8bit produces a non-power-of-2 force the

258 mantissa-rounding-up branch (exp_byte += 1). Locks down this behavior

260 - before optimization.

261 - """

259 + before optimization."""

260 device = "cuda"

261 num_tokens, hidden_dim, group_size = 4, 7168, 128

262

@@ -330,8 +328,7 @@ def test_per_token_group_quant_fp8_packed_zero_fills_padded_output_q(

328 ):

329 """When output_q is allocated with shape (tma_aligned_mn, k) instead of

330 (mn, k), the kernel must overwrite the padded mn rows with zeros so

333 - callers can use ``torch.empty`` instead of ``torch.zeros``.

334 - """

331 + callers can use ``torch.empty`` instead of ``torch.zeros``."""

332 device = "cuda"

333 group_size = 128

334 torch.manual_seed(42)

tests/kernels/quantization/test_quantized_embedding.py

+1 -2

@@ -1,8 +1,7 @@

1 # SPDX-License-Identifier: Apache-2.0

2 # SPDX-FileCopyrightText: Copyright contributors to the vLLM project

3 """Tests for the Triton dequant-gather kernel used by

4 -``CompressedTensorsEmbeddingWNA16Int`` (quantized embedding lookup).

5 -"""

4 +``CompressedTensorsEmbeddingWNA16Int`` (quantized embedding lookup)."""

5

6 import pytest

7 import torch

tests/kernels/quantization/test_rocm_aiter_grouped_quant.py

+2 -4

@@ -29,8 +29,7 @@ pytestmark = pytest.mark.skipif(

29

30 def test_rocm_aiter_group_fp8_quant_fake_implementation():

31 """Test that the fake implementation is correctly

32 - defined for torch.ops.vllm.rocm_aiter_group_fp8_quant.

33 - """

32 + defined for torch.ops.vllm.rocm_aiter_group_fp8_quant."""

33 # Create test tensors

34 M = 128

35 N = 4096

@@ -50,8 +49,7 @@ def test_rocm_aiter_group_fp8_quant_fake_implementation():

49 def test_rocm_aiter_group_fp8_quant_torch_compile_with_cudagraph():

50 """Test that rocm_aiter_ops.group_fp8_quant

51 with group size 128 can be used with

53 - torch.compile in cudagraph mode.

54 - """

52 + torch.compile in cudagraph mode."""

53 # Create test tensors

54 M = 128

55 N = 4096

tests/kernels/quantization/test_rocm_mxfp4.py

+1 -2

@@ -209,8 +209,7 @@ def test_rocm_aiter_fp4_enablement_follows_env_and_arch(

209 @pytest.mark.parametrize("shape", LLAMA_SHAPES)

210 def test_vllm_quant_dequant_mxfp4_matches_reference_on_large_shapes(shape, dtype):

211 """The public vLLM MXFP4 QDQ helper should stay exact against the torch

212 - reference even on larger Llama-like shapes.

213 - """

212 + reference even on larger Llama-like shapes."""

213 from vllm.model_executor.layers.quantization.utils.mxfp4_utils import (

214 quant_dequant_mxfp4,

215 )

tests/kernels/test_compressor_kv_cache.py

+3 -6

@@ -670,8 +670,7 @@ def test_v41_rope_insert_plain_row(compress_ratio: int, store_fp8: bool):

670 def test_v41_compressor_metadata_maps_tokens_to_their_ring():

671 """The ring group's generic slot mapping is disabled (all PAD), so the

672 builder must map every real token to ``ring_block * capacity + pos %

673 - capacity`` and keep padding tokens at PAD.

674 - """

673 + capacity`` and keep padding tokens at PAD."""

674 from unittest.mock import MagicMock

675

676 from vllm.models.deepseek_v41.compressor import CompressorMetadataBuilder

@@ -1096,8 +1095,7 @@ def test_get_c128_boundary(starts, query_start_loc, expected):

1095 @pytest.mark.parametrize("block_size", [16, 64])

1096 def test_deepseek_v4_attention_quant_cache_roundtrip(num_tokens: int, block_size: int):

1097 """compressed_kv → quantize_and_insert_k_cache → dequantize_and_gather_k_cache

1099 - → compare against original.

1100 - """

1098 + → compare against original."""

1099 HEAD_DIM = 512

1100 NOPE_DIM = 448

1101 HEAD_BYTES = 584 # 448 fp8 + 128 bf16 + 8 uint8 scale

@@ -1309,8 +1307,7 @@ def test_dequantize_and_gather_k_cache(

1307 @pytest.mark.parametrize("block_size", [16, 64])

1308 def test_indexer_quant_cache_roundtrip(num_tokens: int, block_size: int):

1309 """K → indexer_k_quant_and_cache → cp_gather_indexer_k_quant_cache

1312 - → manual dequant → compare against original.

1313 - """

1310 + → manual dequant → compare against original."""

1311 HEAD_DIM = 128

1312 QUANT_BLOCK_SIZE = 128

1313 # cache_stride = head_dim + (head_dim * 4 / quant_block_size) = 128 + 4 = 132

tests/kernels/test_cp_gather_fp8.py

+1 -2

@@ -432,8 +432,7 @@ def test_cp_gather_fp8_with_sequence_starts(gather_seq_lens, seq_starts):

432 )

433 def test_cp_gather_fp8_large_seqlens(seq_lens, block_size):

434 """Correctness test with large sequence lengths matching benchmark

435 - scenarios (8K-128K prefill).

436 - """

435 + scenarios (8K-128K prefill)."""

436 (

437 cache,

438 block_table,

tests/kernels/test_deepseek_v4_cpu_kernels.py

+19 -38

@@ -85,8 +85,7 @@ def _decode_token_regions(

85 cache: torch.Tensor, slot_mapping: torch.Tensor, block_size: int

86 ):

87 """Slice out, per valid token, the raw NoPE-fp8 / RoPE-bf16 / scale byte

88 - regions this layout defines.

89 - """

88 + regions this layout defines."""

89 valid = slot_mapping >= 0

90 idx = valid.nonzero(as_tuple=True)[0]

91 slots = slot_mapping[idx]

@@ -111,8 +110,7 @@ def _assert_cache_parity(cache_fused, cache_ref, slot_mapping, block_size):

110 """NoPE fp8 + the UE8M0 scale byte are a deterministic quant of an

111 un-rotated bf16 value, so they match tightly; the RoPE region is rotated

112 in fp32 before its bf16 store and eager/kernel can round a tie to

114 - opposite sides, so allow <=1 ULP there.

115 - """

113 + opposite sides, so allow <=1 ULP there."""

114 nope_ref, rope_ref, scale_ref = _decode_token_regions(

115 cache_ref, slot_mapping, block_size

116 )

@@ -192,8 +190,7 @@ def _dup_padded_block_cache(

190 num_blocks: int, block_size: int, width: int, pad_elems: int

191 ) -> tuple[torch.Tensor, torch.Tensor]:

192 """Two independently-stored padded caches with identical values, for ops

195 - that mutate the cache in place.

196 - """

193 + that mutate the cache in place."""

194 page_stride = block_size * width + pad_elems

195 template = torch.randn(num_blocks, page_stride, dtype=torch.float32)

196

@@ -210,8 +207,7 @@ def _run_save_partial_states_eager(

207 ):

208 """Reference: state_cache[block_idx, pos_in_block, :W] = kv;

209 state_cache[..., W:] = score + ape[position % compress_ratio]; skip

213 - slot < 0. Mutates state_cache in place, matching the op's contract.

214 - """

210 + slot < 0. Mutates state_cache in place, matching the op's contract."""

211 state_width = kv.shape[-1]

212 block_size = state_cache.shape[1]

213 valid = slot_mapping >= 0

@@ -271,8 +267,7 @@ def _run_compress_eager(

267 ):

268 """Reference for the per-channel-softmax-pooling + RMSNorm + NoPE-UE8M0

269 quant + GPT-J RoPE math writing the main (head_dim=512) fp8_ds_mla

274 - layout.

275 - """

270 + layout."""

271 num_tokens, window = gather_slots.shape

272 state_width = state_cache.shape[-1] // 2

273 state_block_size = state_cache.shape[1]

@@ -449,8 +444,7 @@ def test_compress_norm_rope_store_cpu_matches_eager(

444

445 def test_compress_norm_rope_store_cpu_non_contiguous_kv_cache_row_stride():

446 """The main (output) kv_cache is a separate physical buffer from the

452 - state cache and can independently be a strided view.

453 - """

447 + state cache and can independently be a strided view."""

448 eps = 1e-6

449 compress_ratio, overlap = 128, False

450 num_tokens = 5

@@ -526,8 +520,7 @@ def _run_compress_indexer_eager(

520 """Same softmax-pool + RMSNorm + GPT-J RoPE prologue as the head=512

521 path, but the indexer's cache quantizes all 128 post-RoPE dims as one

522 FP8 block with a single raw fp32 scale per token (matching

529 - _indexer_k_quant_and_cache_cpu's layout), not 7 separate UE8M0 blocks.

530 - """

523 + _indexer_k_quant_and_cache_cpu's layout), not 7 separate UE8M0 blocks."""

524 num_tokens, window = gather_slots.shape

525 state_width = state_cache.shape[-1] // 2

526 state_block_size = state_cache.shape[1]

@@ -655,8 +648,7 @@ def _pack_flash_mla_cache(kv_f32: torch.Tensor, block_size: int):

648 """Quantize kv_f32 ([N, HEAD_DIM]) into the fp8_ds_mla byte layout and

649 return (cache, dequant_ref) -- the exact post-quantization value the

650 kernel's dequant should reproduce. Slot id == row index, packed

658 - contiguously from block 0.

659 - """

651 + contiguously from block 0."""

652 n = kv_f32.shape[0]

653 num_blocks = max(1, -(-n // block_size))

654 cache = _make_cache(num_blocks, block_size)

@@ -714,8 +706,7 @@ def _flash_mla_eager(

706 """MQA attention: gather the (already-dequantized) window+compressed KV

707 rows referenced by *_slots (-1 = invalid/masked), append a learned

708 per-head sink logit with zero value contribution, softmax, weighted

717 - sum.

718 - """

709 + sum."""

710 num_tokens, num_heads, _ = q.shape

711 q_f32 = q.float()

712

@@ -814,8 +805,7 @@ def test_flash_mla_cpu_kernel_matches_eager(

805

806 def test_flash_mla_cpu_kernel_all_invalid_falls_back_to_sink():

807 """When every KV slot is masked out, softmax degenerates to the sink

817 - logit alone and the output must be all-zero (no valid value rows).

818 - """

808 + logit alone and the output must be all-zero (no valid value rows)."""

809 torch.manual_seed(2)

810 dtype = torch.bfloat16

811 num_tokens, num_heads, num_window = 3, 4, 5

@@ -855,8 +845,7 @@ def _qnorm_kv_insert_eager(

845 q, kv, positions, cache, slot_mapping, cos_sin_cache, q_head_padded, eps, bs

846 ):

847 """Reference: per-head weight-free RMSNorm + GPT-J RoPE on Q; GPT-J RoPE

858 - + UE8M0 FP8 quant + paged insert on KV.

859 - """

848 + + UE8M0 FP8 quant + paged insert on KV."""

849 num_quant_blocks = NOPE_DIM // QUANT_BLOCK

850

851 num_tokens_full, num_heads_q, _ = q.shape

@@ -982,8 +971,7 @@ def test_qnorm_rope_kv_insert_cpu_kernel_matches_eager(

971 @pytest.mark.parametrize("pad", [1, 5])

972 def test_qnorm_rope_kv_insert_cpu_kernel_dp_padding(pad: int):

973 """slot_mapping shorter than q/kv rows: the KV branch must only touch

985 - the first num_tokens rows while Q-norm+RoPE still runs on all rows.

986 - """

974 + the first num_tokens rows while Q-norm+RoPE still runs on all rows."""

975 torch.manual_seed(1)

976 dtype = torch.bfloat16

977 eps = 1e-6

@@ -1044,8 +1032,7 @@ def _inverse_gptj_rope_o_proj_eager(

1032 rope_dim: int,

1033 ) -> torch.Tensor:

1034 """NoPE dims pass through as fp32; the RoPE segment is de-rotated by

1047 - R(-pos) in GPT-J (interleaved even/odd) convention.

1048 - """

1035 + R(-pos) in GPT-J (interleaved even/odd) convention."""

1036 nope_dim = o.shape[-1] - rope_dim

1037 half_dim = rope_dim // 2

1038

@@ -1111,8 +1098,7 @@ def _write_page(

1098 ) -> None:

1099 """Write k_fp8 ([block_size, INDEXER_HEAD_DIM]) and scale ([block_size]

1100 fp32) into one page row: K-region then scale-region, matching

1114 - compress_norm_rope_store_indexer_cpu's layout.

1115 - """

1101 + compress_norm_rope_store_indexer_cpu's layout."""

1102 k_region = page_row[: block_size * INDEXER_HEAD_DIM]

1103 k_region.copy_(k_fp8.view(torch.uint8).flatten())

1104 scale_region = page_row[block_size * INDEXER_HEAD_DIM :].view(torch.float32)

@@ -1148,8 +1134,7 @@ def _build_paged_mqa_logits_batch(

1134 ):

1135 """Fill pages (a pre-sized [n_pages, buf_width] uint8 tensor, possibly

1136 non-contiguous in dim 0) with deterministic-per-seed random K/scale

1151 - data.

1152 - """

1137 + data."""

1138 torch.manual_seed(seed)

1139 batch_size = len(seq_lens)

1140 max_seq_len = max(seq_lens)

@@ -1215,8 +1200,7 @@ def test_fp8_paged_mqa_logits_cpu_non_contiguous_kv_cache_row_stride():

1200 """The paged K-cache in production is a per-layer view into a shared

1201 multi-layer allocation, so its page stride generally exceeds the page

1202 byte width -- check the kernel reads a padded/strided buffer identically

1218 - to an equivalent tightly-packed one.

1219 - """

1203 + to an equivalent tightly-packed one."""

1204 block_size = 64

1205 num_heads = 4

1206 seq_lens = [1, block_size + 7, 3 * block_size]

@@ -1342,8 +1326,7 @@ def test_fused_indexer_q_rope_quant_dispatches_to_cpu_kernel():

1326 CPU op on a CPU tensor, not silently fall through to the Triton kernel

1327 (which would also "work" under triton-cpu, masking a dispatch

1328 regression). The dispatcher does its CPU import inside the CPU branch,

1345 - so the patch target is the source module.

1346 - """

1329 + so the patch target is the source module."""

1330 torch.manual_seed(0)

1331 num_tokens = 1

1332

@@ -1410,8 +1393,7 @@ def _hc_head_torch(

1393 ) -> torch.Tensor:

1394 """Eager reference for hc_head_fused_cpu: round-trips the RMSNorm

1395 output through bf16 before the linear, matching the kernel's bf16

1413 - arithmetic.

1414 - """

1396 + arithmetic."""

1397 hc_mult, hidden_size = hidden_states.shape[-2:]

1398 outer_shape = hidden_states.shape[:-2]

1399 hs_flat = hidden_states.reshape(-1, hc_mult, hidden_size)

@@ -1493,8 +1475,7 @@ def test_mhc_post_cpu(num_tokens: int, hidden_size: int):

1475 @pytest.mark.parametrize("hidden_size", [4096, 7168])

1476 def test_hc_head_cpu(num_tokens: int, hidden_size: int):

1477 """hc_head_fused_cpu vs. the _hc_head_torch eager reference it

1496 - replaces.

1497 - """

1478 + replaces."""

1479 torch.set_default_device(DEVICE)

1480 set_random_seed(0)

1481 hc_mult = 4

tests/kernels/test_engram.py

+3 -6

@@ -383,8 +383,7 @@ def _hash_ids(requests, tables_meta, block_table, cache=None, capture=False):

383 """Run the op on a batch of (token_ids, start, window) requests; `window`

384 is the runner's [depth] lookback for that request (None = all unknown).

385 Token IDs 14 and 22 stand in for image tokens that break n-grams.

386 - Returns one [len(token_ids), ...] hash tensor per request.

387 - """

386 + Returns one [len(token_ids), ...] hash tensor per request."""

387 token_map, multipliers, primes, offsets, block_size = tables_meta

388 depth = multipliers.shape[1] - 1

389

@@ -443,8 +442,7 @@ def test_lookback_window_reproduces_single_instance(runner, capture):

442 V2 supplies every lookback from its device token history and needs no

443 slot cache. V1 supplies prompt positions only and reads generated

444 positions from the slot cache it fills itself. Without any window the

446 - first decode token hashes stale slots (negative control).

447 - """

445 + first decode token hashes stale slots (negative control)."""

446 block_size = 4

447 block_table = torch.tensor(

448 [[3, 1, 5, 0], [2, 4, 6, 7]], dtype=torch.int32, device="cuda"

@@ -710,8 +708,7 @@ def test_engram_lookup_reuses_jit_across_token_shapes():

708 def test_engram_lookup_matches_torch(cpu_offload, background, num_tokens):

709 """The fused gather must be bit-exact with the torch dequant path it

710 replaces, from HBM and from pinned host memory alike, and must contribute

713 - zeros for rows another TP rank owns.

714 - """

711 + zeros for rows another TP rank owns."""

712 layer = _make_embedding(cpu_offload)

713 cols, rows = 24, layer.part_num_embeddings

714 ids = torch.randint(0, rows, (num_tokens, cols), dtype=torch.int32, device="cuda")

tests/kernels/test_fla_layernorm_guard.py

+1 -2

@@ -336,8 +336,7 @@ def test_layer_norm_rows_per_block(

336 @torch.inference_mode()

337 def test_strided_input(dtype: torch.dtype) -> None:

338 """Test that the kernel handles non-contiguous (strided)

339 - inputs correctly.

340 - """

339 + inputs correctly."""

340 set_random_seed(42)

341 device = torch.device(DEVICE)

342 num_tokens = 128

tests/kernels/test_fp32_router_gemm.py

+1 -2

@@ -97,8 +97,7 @@ def test_topk_routing_consistency(num_tokens: int, hidden_dim: int, num_experts:

97 """The gate feeds top-k expert selection: the kernel's top-8 must match

98 an fp64 reference's top-8 per token (ties tolerated). This is the

99 business-level correctness of the router — numeric error only matters

100 - if it flips the argsort.

101 - """

100 + if it flips the argsort."""

101 _requires_sm90()

102 top_k = 8

103 device = torch.device("cuda")

tests/kernels/test_fused_deepseek_v32_norm_rope.py

+1 -2

@@ -854,8 +854,7 @@ def test_fused_q_no_indexer(num_tokens: int):

854 @pytest.mark.parametrize("has_indexer", [True, False])

855 def test_fused_q_bf16_query(num_tokens: int, has_indexer: bool):

856 """bf16-query path (FlashMLA sparse, SM90/SM100): only the RoPE'd q_pe is

857 - produced (bf16, unquantized); ql_nope is consumed directly by the caller.

858 - """

857 + produced (bf16, unquantized); ql_nope is consumed directly by the caller."""

858 torch.manual_seed(6)

859 dev = "cuda"

860 max_pos = 8192

tests/kernels/test_fused_deepseek_v4_qnorm_rope_kv_insert.py

+2 -4

@@ -176,8 +176,7 @@ def _call_fused(

176

177 def _as_stored_fp8(t: torch.Tensor) -> torch.Tensor:

178 """Reinterpret a float8_e4m3fn-typed kernel output under the real (FNUZ on

179 - gfx942) encoding the kernel actually wrote, without touching the bytes.

180 - """

179 + gfx942) encoding the kernel actually wrote, without touching the bytes."""

180 return t.contiguous().view(torch.uint8).view(FP8_STORE_DTYPE)

181

182

@@ -713,8 +712,7 @@ def test_q_interleaved_is_orthogonal_to_norm_and_rope(

712 @pytest.mark.parametrize("block_size", [16, 64])

713 def test_kv_path_with_dp_padding(num_tokens: int, pad: int, block_size: int):

714 """slot_mapping.size(0) < q.size(0): the kernel must skip padded

716 - tokens in the KV branch while still processing Q on all rows.

717 - """

715 + tokens in the KV branch while still processing Q on all rows."""

716 torch.manual_seed(3)

717 device = "cuda"

718 dtype = torch.bfloat16

tests/kernels/test_fused_indexer_q_rope_quant.py

+1 -2

@@ -574,8 +574,7 @@ def test_indexer_k_store_roundtrips_through_rocm_gather(block_size, compress_rat

574 def test_fused_indexer_q_rope_quant_writes_bf16_weights(use_cutedsl):

575 """The MXFP4 path can emit the per-head weights in bf16 (what DeepGEMM's

576 sparse MQA-logits kernels take) instead of fp32; the values are the fp32

577 - result rounded once, so the scoring kernel needs no cast.

578 - """

577 + result rounded once, so the scoring kernel needs no cast."""

578 if use_cutedsl and not has_cutedsl():

579 pytest.skip("cutedsl (cutlass) not installed")

580

tests/kernels/test_fused_inv_rope_fp8_quant.py

+3 -6

@@ -446,8 +446,7 @@ def test_output_strides(num_tokens, num_heads, n_groups):

446 def test_per_group_contiguity(num_tokens):

447 """FP8 per-group slices must be contiguous. Scale per-group slices

448 are column-major (T-stride=1) — not row-major contiguous, which is

449 - correct for TMA loads.

450 - """

449 + correct for TMA loads."""

450 num_heads, n_groups = 64, 8

451 heads_per_group = num_heads // n_groups

452 max_pos = 4096

@@ -513,8 +512,7 @@ def test_scales_are_power_of_two():

512 def test_nope_dims_unchanged():

513 """Nope dimensions (first 448 per head) should only be quantized,

514 not rotated. Verify by dequantizing and comparing against

516 - quantize-only reference (no RoPE).

517 - """

515 + quantize-only reference (no RoPE)."""

516 num_tokens, num_heads, n_groups = 16, 64, 8

517 heads_per_group = num_heads // n_groups

518 max_pos = 4096

@@ -905,8 +903,7 @@ def test_einsum_end_to_end(num_tokens, num_heads, n_groups):

903 @torch.inference_mode()

904 def test_with_real_deepseek_v4_rope(num_tokens, default_vllm_config):

905 """Test with real DeepseekV4ScalingRotaryEmbedding (GPT-J style,

908 - mscale=0, YaRN scaling) matching the production config.

909 - """

906 + mscale=0, YaRN scaling) matching the production config."""

907 num_heads = 64

908 n_groups = 8

909 heads_per_group = num_heads // n_groups

tests/kernels/test_mhc_kernels.py

+2 -4

@@ -1372,8 +1372,7 @@ def _patch_first_rank_pp_group(monkeypatch):

1372

1373 def test_deepseek_v4_mhc_broadcast_finalize_sums_hc_streams(monkeypatch):

1374 """First finalize (at the end of load_weights) allocates

1375 - hc_attn_fn_broadcast as hc_attn_fn summed over hc streams.

1376 - """

1375 + hc_attn_fn_broadcast as hc_attn_fn summed over hc streams."""

1376 _patch_first_rank_pp_group(monkeypatch)

1377 layer = _make_mhc_decoder_layer(hc_mult=2, hidden_size=8)

1378 model = SimpleNamespace(start_layer=0, end_layer=1, layers=[layer])

@@ -1388,8 +1387,7 @@ def test_deepseek_v4_mhc_broadcast_finalize_sums_hc_streams(monkeypatch):

1387 def test_deepseek_v4_mhc_broadcast_refit_refreshes_in_place(monkeypatch):

1388 """Re-finalizing after a weight refit must copy into the existing

1389 broadcast tensor so its address stays stable for captured CUDA graphs,

1391 - while picking up the new hc_attn_fn values.

1392 - """

1390 + while picking up the new hc_attn_fn values."""

1391 _patch_first_rank_pp_group(monkeypatch)

1392 layer = _make_mhc_decoder_layer(hc_mult=2, hidden_size=8)

1393 model = SimpleNamespace(start_layer=0, end_layer=1, layers=[layer])

tests/kernels/test_minimax_m3_amd_ops.py

+4 -8

@@ -403,8 +403,7 @@ def test_mxfp8_linear_emulation_bf16_at_load(

403 ):

404 """EmulationMxfp8LinearKernel load-time BF16 dequant (default) and the

405 ``VLLM_MXFP8_EMULATION_DEQUANT_AT_LOAD=0`` per-step fallback must produce the

406 - same result; the dtype-match (BF16/FP16 activations) must also hold.

407 - """

406 + same result; the dtype-match (BF16/FP16 activations) must also hold."""

407 from vllm.model_executor.kernels.linear.mxfp8.emulation import (

408 EmulationMxfp8LinearKernel,

409 )

@@ -518,8 +517,7 @@ def test_mxfp8_rocm_native_unaligned_k_dequantizes_at_load(shape):

517 # produced EP garbage).

518 def _capture_expert_mask(expert_mask, *, global_num_experts):

519 """Drive the real ``AiterMxfp8Experts.apply`` mask branch and capture the

521 - ``expert_mask`` it forwards to ``rocm_aiter_ops.fused_moe``.

522 - """

520 + ``expert_mask`` it forwards to ``rocm_aiter_ops.fused_moe``."""

521 from types import SimpleNamespace

522 from unittest import mock

523

@@ -570,8 +568,7 @@ def _capture_expert_mask(expert_mask, *, global_num_experts):

568 @pytest.mark.skipif(not current_platform.is_rocm(), reason="ROCm only")

569 def test_aiter_mxfp8_apply_forwards_expert_mask_unchanged():

570 """AiterMxfp8Experts.apply forwards the precomputed 0/1 mask to aiter as-is

573 - (no re-derivation that would collapse an already-0/1 mask to all-ones).

574 - """

571 + (no re-derivation that would collapse an already-0/1 mask to all-ones)."""

572 # 0/1 local-expert mask over global ids + trailing sentinel (rank owns 0..3).

573 ep_mask = torch.tensor(

574 [1, 1, 1, 1, 0, 0, 0, 0, 0], dtype=torch.int32, device=DEVICE

@@ -585,8 +582,7 @@ def test_aiter_mxfp8_apply_forwards_expert_mask_unchanged():

582 def test_routed_experts_expert_map_delegates_to_kernel():

583 """RoutedExperts.expert_map returns the 0/1 mask only for kernels that set

584 ``consumes_expert_mask`` (AITER), and the canonical -1 map otherwise -- keyed

588 - on the resolved kernel, not the global aiter switch.

589 - """

585 + on the resolved kernel, not the global aiter switch."""

586 from types import SimpleNamespace

587

588 from vllm.model_executor.layers.fused_moe.routed_experts import RoutedExperts

tests/kernels/test_shuffle_rows.py

+2 -4

@@ -17,8 +17,7 @@ from vllm.platforms import current_platform

17 @pytest.mark.parametrize("dtype", [torch.float16, torch.bfloat16, torch.float32])

18 def test_shuffle_rows_basic(num_tokens: int, hidden_size: int, dtype: torch.dtype):

19 """Test basic functionality of shuffle_rows with various tensor sizes and

20 - dtypes.

21 - """

20 + dtypes."""

21 if not current_platform.is_cuda():

22 pytest.skip("shuffle_rows requires CUDA")

23

@@ -73,8 +72,7 @@ def test_shuffle_rows_permutation(

72 @pytest.mark.parametrize("hidden_size", [256, 512])

73 def test_shuffle_rows_expansion(num_tokens: int, hidden_size: int):

74 """Test shuffle_rows with expansion (more output tokens than input

76 - tokens).

77 - """

75 + tokens)."""

76 if not current_platform.is_cuda():

77 pytest.skip("shuffle_rows requires CUDA")

78

tests/kernels/test_top_k_per_row.py

+2 -4

@@ -1740,8 +1740,7 @@ def test_sparse_indexer_decode_topk_explicit_backends(

1740 ) -> None:

1741 """Every explicit sparse-indexer decode top-k backend must restrict each

1742 row to its seq_len (dirty data past the end must never be selected) and

1743 - match torch.topk on the valid region.

1744 - """

1743 + match torch.topk on the valid region."""

1744 from vllm.config import VllmConfig, set_current_vllm_config

1745 from vllm.model_executor.layers.indexer_topk import SparseIndexerTopk

1746

@@ -1812,8 +1811,7 @@ def test_sparse_indexer_decode_topk_short_seq_lens(

1811 backend: str, workspace_init

1812 ) -> None:

1813 """seq_len < next_n: derived per-row ends must clamp at 0 (the reference

1815 - kernels do max(0, ...)); negative ends are OOB for the ragged kernels.

1816 - """

1814 + kernels do max(0, ...)); negative ends are OOB for the ragged kernels."""

1815 from vllm.config import VllmConfig, set_current_vllm_config

1816 from vllm.model_executor.layers.indexer_topk import SparseIndexerTopk

1817

tests/lora/test_default_mm_loras.py

+1 -2

@@ -132,8 +132,7 @@ def test_default_mm_lora_fails_with_overridden_lora_request(

132 monkeypatch: pytest.MonkeyPatch,

133 ):

134 """Ensure that if the lora_request conflicts with default_mm_loras,

135 - we use the lora_request.

136 - """

135 + we use the lora_request."""

136 run_test(

137 vllm_runner,

138 audio_assets,

tests/lora/test_fused_moe_lora_kernel.py

+2 -4

@@ -1123,8 +1123,7 @@ def test_fused_moe_lora_kernel_one_shot_early_exit(trigger, device):

1123 @pytest.mark.parametrize("device", DEVICES)

1124 def test_fused_moe_lora_kernel_zero_grid_no_crash(device):

1125 """num_active_loras=0 (or num_slices=0) would otherwise launch a grid

1126 - with a zero dimension. one-shot wrapper must short-circuit before launch.

1127 - """

1126 + with a zero dimension. one-shot wrapper must short-circuit before launch."""

1127 torch.set_default_device(device)

1128 set_random_seed(0)

1129 num_tokens, top_k, E, max_loras, R, K, N = 8, 2, 8, 4, 16, 1024, 1024

@@ -1186,8 +1185,7 @@ def test_fused_moe_lora_kernel_zero_grid_no_crash(device):

1185 def test_fused_moe_lora_kernel_rejects_bad_block_size_m(device):

1186 """one-shot must surface a clear assertion when shrink_block_size_m is not

1187 a power of 2 / less than 16, instead of the cryptic Triton compile

1189 - failure (`arange's range must be a power of 2`).

1190 - """

1188 + failure (`arange's range must be a power of 2`)."""

1189 torch.set_default_device(device)

1190 set_random_seed(0)

1191 num_tokens, top_k, E, max_loras, R, K, N = 32, 2, 8, 4, 16, 1024, 1024

tests/lora/test_moe_lora_align_sum.py

+2 -4

@@ -249,8 +249,7 @@ def _build_and_run_align(

249 )

250 def test_moe_lora_align_block_size_mixed_base_and_lora(max_loras):

251 """Regression test for issue #32235: real LoRA slot must not be skipped

252 - when ``active_lora_ids`` has -1 at position 0.

253 - """

252 + when ``active_lora_ids`` has -1 at position 0."""

253 out = _build_and_run_align(

254 num_lora_tokens=8, num_base_tokens=8, max_loras=max_loras

255 )

@@ -289,8 +288,7 @@ def test_moe_lora_align_block_size_disabled_adapter_untouched():

288 kernels. Pins the invariant protected by the ``adapter_enabled`` guard

289 in ``lora_count_and_sort_expert_tokens_kernel``: without it the sort

290 kernel reads uninitialized ``token_mask`` values for disabled slots and

292 - pollutes ``sorted_token_ids`` / ``cumsum_buffer``.

293 - """

291 + pollutes ``sorted_token_ids`` / ``cumsum_buffer``."""

292 max_loras = 1

293 out = _build_and_run_align(

294 num_lora_tokens=16,

tests/lora/test_moe_lora_ep_load.py

+2 -4

@@ -19,8 +19,7 @@ NON_EXPERT_MODULES = ("q_proj", "k_proj", "v_proj", "o_proj", "gate")

19

20 def _expected_lora_modules() -> set[str]:

21 """Replicate the set ``WorkerLoRAManager._load_adapter`` would build

22 - from this model's ``packed_modules_mapping``.

23 - """

22 + from this model's ``packed_modules_mapping``."""

23 expected: set[str] = set(NON_EXPERT_MODULES)

24 for expert in range(GLOBAL_NUM_EXPERTS):

25 for proj in EXPERT_PROJECTIONS:

@@ -43,8 +42,7 @@ def _load(lora_dir, peft_helper, *, moe_ep_spec, lora_id):

42 def test_moe_lora_ep2_real_qwen3moe(qwen3moe_lora_files, ep_rank):

43 """ep_size=2 against the real Qwen3-MoE adapter: each rank's loaded

44 LoRA has the right size, the right expert membership, and the

46 - right tensor values.

47 - """

45 + right tensor values."""

46 peft_helper = PEFTHelper.from_local_dir(

47 qwen3moe_lora_files, max_position_embeddings=4096

48 )

tests/lora/test_punica_ops.py

+2 -4

@@ -45,8 +45,7 @@ def _bgmv_shrink(

45 inputs, lora_weight, output, seq_len_tensor, lora_indices, scaling=1.0

46 ):

47 """Memory-efficient shrink reference: per-LoRA matmul loop.

48 - output[mask] = scaling * inputs[mask] @ weight.T

49 - """

48 + output[mask] = scaling * inputs[mask] @ weight.T"""

49 exploded = torch.repeat_interleave(lora_indices, seq_len_tensor)

50 for lid in exploded.unique():

51 if lid < 0:

@@ -67,8 +66,7 @@ def _bgmv_expand(

66 add_inputs=False,

67 ):

68 """Memory-efficient expand reference: per-LoRA matmul loop.

70 - output[mask, offset:offset+n] (+)= inputs[mask] @ weight.T

71 - """

69 + output[mask, offset:offset+n] (+)= inputs[mask] @ weight.T"""

70 exploded = torch.repeat_interleave(lora_indices, seq_len_tensor)

71 for lid in exploded.unique():

72 if lid < 0:

tests/lora/test_punica_ops_fp8.py

+2 -4

@@ -877,8 +877,7 @@ def test_lora_shrink_fp8(

877 quant_mode: str,

878 ):

879 """Test FP8 shrink kernel with per-tensor, per-channel, and block-wise

880 - quantization, comparing against the bf16 baseline.

881 - """

880 + quantization, comparing against the bf16 baseline."""

881 torch.set_default_device(device)

882 set_random_seed(seed)

883

@@ -933,8 +932,7 @@ def test_lora_expand_fp8(

932 quant_mode: str,

933 ):

934 """Test FP8 expand kernel with per-tensor, per-channel, and block-wise

936 - quantization, comparing against the bf16 baseline.

937 - """

935 + quantization, comparing against the bf16 baseline."""

936 torch.set_default_device(device)

937 set_random_seed(seed)

938

tests/model_executor/layers/test_fused_shared_expert.py

+2 -4

@@ -1319,8 +1319,7 @@ def test_fp8_shared_expert_fse_allows_symmetric_ignored_layers() -> None:

1319

1320 def test_fp8_shared_expert_fse_expands_packed_projections() -> None:

1321 """Both shards of a fused projection ignored is symmetric within that

1322 - projection, so the mismatch reported is against the routed experts.

1323 - """

1322 + projection, so the mismatch reported is against the routed experts."""

1323 quant_config = _fp8_config(

1324 ignored_layers=[

1325 "model.layers.0.mlp.shared_experts.gate_proj",

@@ -1344,8 +1343,7 @@ def test_fp8_shared_expert_fse_expands_packed_projections() -> None:

1343

1344 def test_fp8_shared_expert_fse_propagates_partial_shard_exclusion() -> None:

1345 """Half a fused projection excluded is rejected by `is_layer_skipped`

1347 - itself, exactly as it is in `Fp8Config.get_quant_method`.

1348 - """

1346 + itself, exactly as it is in `Fp8Config.get_quant_method`."""

1347 quant_config = _fp8_config(

1348 ignored_layers=["model.layers.0.mlp.shared_experts.gate_proj"]

1349 )

tests/model_executor/layers/test_mla_short_prefill_indexer.py

+12 -24

@@ -346,8 +346,7 @@ def test_deepseek_v32_dispatches_selected_mha(

346 def test_select_candidate_blocks_tolerates_empty_rows():

347 """Full-cudagraph decode pads the batch with seq_len-0 rows. The newest

348 block pin must not index -1 for them (device-side assert); they select

349 - no candidate blocks and real rows still pin their newest block.

350 - """

349 + no candidate blocks and real rows still pin their newest block."""

350 block_size, topk_blocks = 8, 3

351 logits = torch.zeros(3, 64, device="cuda")

352 logits[0, 3] = 5.0 # block 0 scores highest for row 0

@@ -458,8 +457,7 @@ def test_candidate_kernels_preserve_packed_bounds_and_padding(

457

458 def _make_candidates(lens: torch.Tensor, num_candidates: int, cbk: int):

459 """Production-faithful candidates: top min(avail, K) blocks per row —

461 - full coverage for lens <= K*cbk, a fixed-size subset beyond that.

462 - """

460 + full coverage for lens <= K*cbk, a fixed-size subset beyond that."""

461 avail = ((lens.cpu() + cbk - 1) // cbk).tolist()

462 rows = []

463 for a in avail:

@@ -495,8 +493,7 @@ def _candidate_token_mask(

493

494 def _reference_topk(masked: torch.Tensor, topk: int) -> torch.Tensor:

495 """Plain ``torch.topk`` over -inf-masked logits; slots past a row's valid

498 - count are -1 (the indexer's "no token" sentinel).

499 - """

496 + count are -1 (the indexer's "no token" sentinel)."""

497 values, indices = masked.topk(topk, dim=1)

498 return torch.where(values > float("-inf"), indices.int(), -1)

499

@@ -509,8 +506,7 @@ def _assert_same_selection(

506 ) -> None:

507 """Both top-k index sets must pick the same multiset of logit values per

508 row (-1 padding must match too). Indices themselves may differ: the

512 - top-k kernels break ties among equal values in nondeterministic order.

513 - """

509 + top-k kernels break ties among equal values in nondeterministic order."""

510 for r in range(logits.shape[0]):

511 base = int(row_starts[r]) if row_starts is not None else 0

512 g, e = got[r][got[r] >= 0], expected[r][expected[r] >= 0]

@@ -522,8 +518,7 @@ def _assert_same_selection(

518

519 def _sparse_scratch(rows: int, num_sparse: int, topk: int) -> dict:

520 """Caller-owned buffers the sparse path writes into (the metadata builder

525 - owns these in production).

526 - """

521 + owns these in production)."""

522 i32 = dict(dtype=torch.int32, device="cuda")

523 return dict(

524 sparse_indices=torch.zeros(rows, num_sparse, **i32),

@@ -538,8 +533,7 @@ def _expand_candidates_reference(

533 """PyTorch reference for `candidate_blocks_to_sparse_indices`: expand

534 each candidate block into ``cbk // sbk`` sparse blocks, keep those inside

535 ``[0, ke - ks)``, sort ascending, anchor at ``ks // sbk``, pad by

541 - repeating the last valid block, and count the valid sparse columns.

542 - """

536 + repeating the last valid block, and count the valid sparse columns."""

537 ratio = cbk // sbk

538 rows, num_candidates = candidates.shape

539 num_sparse = num_candidates * ratio

@@ -565,8 +559,7 @@ def _expand_candidates_reference(

559 @pytest.mark.skipif(not torch.cuda.is_available(), reason="Triton kernel")

560 def test_candidate_blocks_to_sparse_indices_math():

561 """Candidate->sparse-block expansion: ks anchoring, range filtering,

568 - repeat-last padding, and the per-row valid column count (``end``).

569 - """

562 + repeat-last padding, and the per-row valid column count (``end``)."""

563 from vllm.model_executor.kernels.attention.dsa.sparse_mqa_logits import (

564 candidate_blocks_to_sparse_indices,

565 )

@@ -618,8 +611,7 @@ def test_candidate_blocks_to_sparse_indices_math():

611 def test_candidate_blocks_to_sparse_indices_matches_reference(cbk: int, sbk: int):

612 """The Triton expansion equals the PyTorch reference on production-shaped

613 random inputs, including strided (column-sliced) candidate rows and

621 - strided output buffers.

622 - """

614 + strided output buffers."""

615 from vllm.model_executor.kernels.attention.dsa.sparse_mqa_logits import (

616 candidate_blocks_to_sparse_indices,

617 )

@@ -689,8 +681,7 @@ def test_sparse_mqa_logits_prefill_matches_dense_masked(capture: bool):

681 ``torch.topk`` over dense bf16 logits masked to the candidates (the two

682 DeepGEMM kernels agree bitwise on candidate columns), incl. under CUDA

683 graph capture, and a later layer reusing the step's expansion/schedule

692 - selects the same values.

693 - """

684 + selects the same values."""

685 deep_gemm = _skip_unless_sm100_sparse_kernels()

686 from vllm.model_executor.kernels.attention.dsa.sparse_mqa_logits import (

687 sparse_mqa_logits_prefill_chunk,

@@ -780,8 +771,7 @@ def test_sparse_mqa_logits_prefill_matches_dense_masked(capture: bool):

771

772 def _build_fp4_paged_cache(req_lens: torch.Tensor, page_kv: int, head_dim: int):

773 """MXFP4 paged cache in DeepGEMM's fused layout: per page, all tokens'

783 - fp4 data first, then their packed UE8M0 scales.

784 - """

774 + fp4 data first, then their packed UE8M0 scales."""

775 device = req_lens.device

776 pages_per = (req_lens + page_kv - 1) // page_kv

777 num_pages, max_pages = int(pages_per.sum()), int(pages_per.max())

@@ -817,8 +807,7 @@ def test_sparse_mqa_logits_paged_matches_dense_masked(next_n: int, capture: bool

807 """Paged sparse-MQA decode selects exactly the values ``torch.topk`` picks

808 from the dense bf16 masked logits, incl. MTP rows (next_n > 1 flattened to

809 one query row each), CUDA graph capture/replay and schedule reuse by a

820 - later layer.

821 - """

810 + later layer."""

811 deep_gemm = _skip_unless_sm100_sparse_kernels()

812 from vllm.model_executor.kernels.attention.dsa.sparse_mqa_logits import (

813 sparse_mqa_logits_paged_decode,

@@ -927,8 +916,7 @@ def test_sparse_mqa_logits_paged_matches_dense_masked(next_n: int, capture: bool

916

917 def test_sparse_mqa_logits_accuracy_vs_torch():

918 """Sparse-kernel numerics vs a PyTorch fp32 reference on the dequantized

930 - MXFP4 inputs.

931 - """

919 + MXFP4 inputs."""

920 _skip_unless_sm100_sparse_kernels()

921 from deep_gemm.utils import cast_back_from_fp4

922

tests/model_executor/model_loader/test_ep_weight_filter.py

+3 -6

@@ -232,8 +232,7 @@ class TestShouldSkipWeight:

232

233 class TestSafetensorsWeightsIteratorWithEpFilter:

234 """Verify that EP filtering produces a strict subset of unfiltered loading

235 - and that all expected dense + local expert weights are present.

236 - """

235 + and that all expected dense + local expert weights are present."""

236

237 @pytest.fixture(scope="class")

238 def gpt2_files(self):

@@ -263,8 +262,7 @@ class TestSafetensorsWeightsIteratorWithEpFilter:

262

263 def test_empty_filter_skips_experts_only(self, gpt2_files):

264 """GPT-2 has no expert weights, so even an empty local_expert_ids

266 - set should return all weights (all are dense).

267 - """

265 + set should return all weights (all are dense)."""

266 all_weights = dict(safetensors_weights_iterator(gpt2_files, False))

267 filtered_weights = dict(

268 safetensors_weights_iterator(gpt2_files, False, local_expert_ids=set())

@@ -275,8 +273,7 @@ class TestSafetensorsWeightsIteratorWithEpFilter:

273

274 class TestEpFilterOnSyntheticMoeWeights:

275 """Create synthetic safetensors files with expert-like naming and verify

278 - that the filter correctly skips non-local experts.

279 - """

276 + that the filter correctly skips non-local experts."""

277

278 @pytest.fixture

279 def synthetic_moe_files(self, tmp_path):

tests/model_executor/model_loader/test_reload.py

+3 -6

@@ -350,8 +350,7 @@ def test_marlin_post_load_preserves_runtime_tensor_addresses(monkeypatch, dist_i

350 @pytest.mark.parametrize("variant", ["fp8", "mxfp8", "nvfp4"])

351 def test_marlin_prepare_layer_preserves_workspace_address(monkeypatch, variant):

352 """The Marlin fallback prepare_* functions rerun on weight reload and must

353 - reuse the workspace storage whose address captured CUDA graphs hold.

354 - """

353 + reuse the workspace storage whose address captured CUDA graphs hold."""

354 from vllm import _custom_ops as ops

355 from vllm.model_executor.layers.quantization.utils import (

356 marlin_utils,

@@ -429,8 +428,7 @@ def test_marlin_prepare_layer_preserves_workspace_address(monkeypatch, variant):

428

429 def test_marlin_make_workspace_new_rejects_incompatible_existing(monkeypatch):

430 """An incompatible existing workspace means the address captured by CUDA

432 - graphs is already unusable; allocating a replacement would hide that.

433 - """

431 + graphs is already unusable; allocating a replacement would hide that."""

432 from vllm.model_executor.layers.quantization.utils import marlin_utils

433

434 monkeypatch.setattr(marlin_utils, "num_compute_units", lambda _: 4)

@@ -611,8 +609,7 @@ class _RecordingQuantMethod(QuantizeMethodBase):

609 class _LateBiasLayer(torch.nn.Module):

610 """Mimics an online-quantized linear: `weight` is created on meta by

611 `create_weights()`, which wraps the loaders, and the linear base registers

614 - `bias` afterwards.

615 - """

612 + `bias` afterwards."""

613

614 def __init__(self, quant_method):

615 super().__init__()

tests/model_executor/test_deep_gemm_warmup.py

+2 -4

@@ -42,8 +42,7 @@ def _run_warmup(monkeypatch, layers) -> list[dict]:

42 @pytest.mark.parametrize("scale_name", ["weight_scale", "weight_scale_inv"])

43 def test_registered_deep_gemm_layer_is_warmed_up(monkeypatch, scale_name) -> None:

44 """Any layer stamped by the DeepGEMM kernel is warmed, regardless of the

45 - quantization method that owns it or the name of its scale parameter.

46 - """

45 + quantization method that owns it or the name of its scale parameter."""

46 layer = _block_fp8_layer(256, 128, scale_name)

47

48 calls = _run_warmup(monkeypatch, [layer])

@@ -55,8 +54,7 @@ def test_registered_deep_gemm_layer_is_warmed_up(monkeypatch, scale_name) -> Non

54

55 def test_n_multiple_of_64_matches_kernel_selection(monkeypatch) -> None:

56 """DeepGEMM accepts N % 64 == 0 (e.g. DeepSeek kv_a_proj_with_mqa, N=576),

58 - so warmup must not require N % 128 == 0.

59 - """

57 + so warmup must not require N % 128 == 0."""

58 assert len(_run_warmup(monkeypatch, [_block_fp8_layer(576, 256)])) == 1

59

60

tests/model_executor/test_routed_experts_capture.py

+1 -2

@@ -326,8 +326,7 @@ def test_routed_experts_capturer_dp_unexpected_batch_raises():

326

327 def test_routed_experts_attention_group_is_shared_and_fail_closed():

328 """Both sides key routing data by this gid, so it must skip non-full-attention

329 - groups rather than defaulting to 0, and fail closed when none exists.

330 - """

329 + groups rather than defaulting to 0, and fail closed when none exists."""

330 common = dict(num_kv_heads=1, head_size=1, dtype=torch.float32)

331 config = SimpleNamespace(

332 kv_cache_groups=[

tests/model_executor/test_weight_utils.py

+8 -16

@@ -57,8 +57,7 @@ class TestMaybeRemapKvScaleName:

57

58 def test_qkv_proj_k_scale(self):

59 """Qwen3-MoE / llm-compressor format: qkv_proj.k_scale -> attn.k_scale

60 - Regression test for https://github.com/vllm-project/vllm/issues/25047

61 - """

60 + Regression test for https://github.com/vllm-project/vllm/issues/25047"""

61 result = maybe_remap_kv_scale_name(

62 "model.layers.0.self_attn.qkv_proj.k_scale", self.PARAMS_DICT

63 )

@@ -66,8 +65,7 @@ class TestMaybeRemapKvScaleName:

65

66 def test_qkv_proj_v_scale(self):

67 """Qwen3-MoE / llm-compressor format: qkv_proj.v_scale -> attn.v_scale

69 - Regression test for https://github.com/vllm-project/vllm/issues/25047

70 - """

68 + Regression test for https://github.com/vllm-project/vllm/issues/25047"""

69 result = maybe_remap_kv_scale_name(

70 "model.layers.0.self_attn.qkv_proj.v_scale", self.PARAMS_DICT

71 )

@@ -110,8 +108,7 @@ class TestMaybeRemapKvScaleName:

108 def test_nvfp4_modelopt_k_proj_k_scale(self):

109 """ModelOpt NVFP4 format (e.g. nvidia/Qwen3-30B-A3B-NVFP4):

110 k_proj.k_scale -> attn.k_scale.

113 - Validates that NVFP4 checkpoints are not broken by this change.

114 - """

111 + Validates that NVFP4 checkpoints are not broken by this change."""

112 result = maybe_remap_kv_scale_name(

113 "model.layers.0.self_attn.k_proj.k_scale", self.PARAMS_DICT

114 )

@@ -120,8 +117,7 @@ class TestMaybeRemapKvScaleName:

117 def test_nvfp4_modelopt_v_proj_v_scale(self):

118 """ModelOpt NVFP4 format (e.g. nvidia/Qwen3-30B-A3B-NVFP4):

119 v_proj.v_scale -> attn.v_scale.

123 - Validates that NVFP4 checkpoints are not broken by this change.

124 - """

120 + Validates that NVFP4 checkpoints are not broken by this change."""

121 result = maybe_remap_kv_scale_name(

122 "model.layers.0.self_attn.v_proj.v_scale", self.PARAMS_DICT

123 )

@@ -129,8 +125,7 @@ class TestMaybeRemapKvScaleName:

125

126 def test_qwen3_vl_moe_qkv_proj_k_scale(self):

127 """Qwen3-VL-MoE uses the same fused qkv_proj naming as Qwen3-MoE.

132 - Regression test for qwen3_vl_moe.py fix (same bug as #25047).

133 - """

128 + Regression test for qwen3_vl_moe.py fix (same bug as #25047)."""

129 result = maybe_remap_kv_scale_name(

130 "model.layers.0.self_attn.qkv_proj.k_scale", self.PARAMS_DICT

131 )

@@ -138,8 +133,7 @@ class TestMaybeRemapKvScaleName:

133

134 def test_qwen3_vl_moe_qkv_proj_v_scale(self):

135 """Qwen3-VL-MoE uses the same fused qkv_proj naming as Qwen3-MoE.

141 - Regression test for qwen3_vl_moe.py fix (same bug as #25047).

142 - """

136 + Regression test for qwen3_vl_moe.py fix (same bug as #25047)."""

137 result = maybe_remap_kv_scale_name(

138 "model.layers.0.self_attn.qkv_proj.v_scale", self.PARAMS_DICT

139 )

@@ -181,8 +175,7 @@ class TestKvCacheScaleMapper:

175 """The `WeightsMapper` returned by `get_cache_scale_mapper` replaces the

176 per-model `maybe_remap_kv_scale_name` calls. It must remap the same set of

177 checkpoint formats (the non-`params_dict`-dependent ones) and be idempotent

184 - so it composes safely with a model's own qkv/gate_up `hf_to_vllm_mapper`.

185 - """

178 + so it composes safely with a model's own qkv/gate_up `hf_to_vllm_mapper`."""

179

180 def _mapper(self):

181 # `get_cache_scale_mapper` does not use `self`; call it on the base

@@ -272,8 +265,7 @@ class TestKvCacheScaleMapper:

265 def test_composes_with_qkv_mapper(self):

266 """Applied together with a model's qkv/gate_up mapper, the regex scale

267 rules run before the substr rename, so scales are normalized to `.attn.`

275 - and regular projections are still fused correctly.

276 - """

268 + and regular projections are still fused correctly."""

269 from vllm.model_executor.models.utils import WeightsMapper

270

271 model_mapper = WeightsMapper(

tests/models/inkling/test_mm_towers.py

+1 -2

@@ -42,8 +42,7 @@ def _assert_close_ulps(

42 compare against the reference's local ulp spacing. ``pre_act`` (the

43 reference pre-activation) adds derivative-propagated slack: near

44 ``gelu(x) ~ 0`` an ulp-level input flip legitimately moves the output by

45 - many of ITS (tiny) ulps, bounded by |gelu'| <= 1.13 times the input ulps.

46 - """

45 + many of ITS (tiny) ulps, bounded by |gelu'| <= 1.13 times the input ulps."""

46 g, r = got.float(), ref.float()

47 tol = torch.clamp(max_ulps * _bf16_spacing(ref), min=atol)

48 if pre_act is not None:

tests/models/kimi_k3/test_amd_latent_moe_runner.py

+1 -2

@@ -123,8 +123,7 @@ def _check_writes_only_its_own_shard(

123 device: torch.device, tp_size: int, rank: int

124 ) -> None:

125 """Two ranks that swapped both weight rows and write offsets still sum to

126 - the right total, so inspect each slice before the final collective.

127 - """

126 + the right total, so inspect each slice before the final collective."""

127 transform = _build_transform(device)

128 runner = _tail_runner(transform, tp_size)

129 group = get_tp_group().device_group

tests/models/kimi_k3/test_aux_attn_res_stream.py

+2 -4

@@ -130,8 +130,7 @@ def test_disabled_reproduces_the_plain_residual_sum(

130

131 def test_taps_the_consumer_layer_when_one_follows(recorder, monkeypatch):

132 """The value the next layer reads is the mixture against *its* weights,

133 - so the tap has to reach forward rather than use the current layer's.

134 - """

133 + so the tap has to reach forward rather than use the current layer's."""

134 _set_last_rank(monkeypatch, True)

135

136 _call(_stub_model(enabled=True), 1, torch.zeros(2), None, torch.zeros(2))

@@ -182,8 +181,7 @@ def test_pending_mlp_output_is_folded_in_rather_than_passed_as_delta(

181 ):

182 """The kernel writes an applied delta back into the prefix in place, which

183 would double-add it into the live residual stream, so the pending output

185 - has to arrive already summed into the prefix with `delta` left None.

186 - """

184 + has to arrive already summed into the prefix with `delta` left None."""

185 _set_last_rank(monkeypatch, True)

186 prefix_sum = torch.tensor([1.0, 2.0])

187 pending = torch.tensor([0.5, 0.25])

tests/models/kimi_k3/test_kda_metadata.py

+1 -2

@@ -727,8 +727,7 @@ def test_cudagraph_capture_metadata_avoids_device_to_host_copy(

727 """`build_for_cudagraph_capture` is not capture-only: a DP rank with nothing

728 scheduled re-stages its FULL-graph metadata through it on every dummy step,

729 so it must not synchronize the device. The host-side draft counts have to

730 - come from `query_start_loc_cpu` and match the device-derived values.

731 - """

730 + come from `query_start_loc_cpu` and match the device-derived values."""

731 num_speculative_tokens = 2

732 batch = BatchSpec(seq_lens=[50, 30], query_lens=[3, 3])

733 device = torch.device("cuda" if torch.cuda.is_available() else "cpu")

tests/models/language/generation/test_gemma.py

+2 -4

@@ -60,8 +60,7 @@ def test_checkpoint_lm_head_can_override_tied_config(monkeypatch) -> None:

60 def test_gemma4_kv_shared_layer_loads_plain_q_proj() -> None:

61 """KV-shared layers have q_proj instead of a packed qkv_proj; their

62 redundant K/V tensors in original checkpoints have no parameter and are

63 - skipped rather than failing the load.

64 - """

63 + skipped rather than failing the load."""

64 model = torch.nn.Module()

65 model.config = SimpleNamespace(num_experts=0)

66 model.start_layer, model.end_layer = 0, 2

@@ -93,8 +92,7 @@ def test_gemma4_kv_shared_layer_loads_plain_q_proj() -> None:

92 @pytest.mark.cpu_test

93 def test_gemma3n_kv_shared_layer_mapper() -> None:

94 """Only non-shared layers pack q/k/v into qkv_proj; KV-shared layers keep

96 - q_proj and drop the redundant K/V tensors original checkpoints ship.

97 - """

95 + q_proj and drop the redundant K/V tensors original checkpoints ship."""

96 config = SimpleNamespace(num_hidden_layers=4, num_kv_shared_layers=2)

97 mapper = Gemma3nTextModel.hf_to_vllm_mapper | _kv_sharing_weights_mapper(config)

98 weights = [

tests/models/language/generation/test_mistral.py

+1 -2

@@ -274,8 +274,7 @@ def test_mistral_function_calling(vllm_runner, model: str, dtype: str) -> None:

274

275 def test_mistral_function_call_nested_json():

276 """Ensure that the function-name regex captures the entire outermost

277 - JSON block, including nested braces.

278 - """

277 + JSON block, including nested braces."""

278

279 # Create a minimal stub tokenizer that provides the few attributes the

280 # parser accesses (`version` and `get_vocab`).

tests/models/language/pooling/test_splade_sparse_pooler.py

+1 -2

@@ -29,8 +29,7 @@ from vllm.v1.worker.gpu.states import RequestState

29 @torch.inference_mode

30 def test_splade_pooler_matches_reference_formula(B, T, H, V):

31 """Ensure SPLADESparsePooler forward() matches the mathematical formula:

32 - log1p(relu(logits)) -> max over sequence length (after masking).

33 - """

32 + log1p(relu(logits)) -> max over sequence length (after masking)."""

33 torch.manual_seed(0)

34

35 # Prepare [B] sequences of shape [T, H]

tests/models/multimodal/generation/test_moondream3_moe.py

+1 -2

@@ -33,8 +33,7 @@ NUM_TOKENS = 3

33

34 class _StubGate(torch.nn.Module):

35 """Router stand-in, so the test does not depend on platform-specific

36 - ``ReplicatedLinear`` weight post-processing.

37 - """

36 + ``ReplicatedLinear`` weight post-processing."""

37

38 def __init__(self, hidden_size: int, num_experts: int):

39 super().__init__()

tests/models/multimodal/generation/test_transformers_audio.py

+1 -2

@@ -81,8 +81,7 @@ def test_transformers_audio_generation(

81 under pytest can cause (possibly flaky) hangs, so they are grouped under

82 the same config. Using VLLM_WORKER_MULTIPROC_METHOD=spawn avoids the

83 deadlock and allows worker processes to terminate cleanly, and release

84 - GPU memory between test runs until the issue is fixed.

85 - """

84 + GPU memory between test runs until the issue is fixed."""

85 # TODO: Remove monkeypatch once

86 # https://github.com/vllm-project/vllm/issues/17676 is fixed.

87 disable_envs_cache()

tests/models/multimodal/generation/vlm_utils/builders.py

+1 -2

@@ -294,8 +294,7 @@ def sample_frames_with_video_metadata(

294

295 def apply_image_size_scaling(image, size: float | tuple[int, int], size_type: SizeType):

296 """Applies a size scaler to one image; this can be an image size factor,

297 - which scales the image while maintaining the aspect ratio

298 - """

297 + which scales the image while maintaining the aspect ratio"""

298 # Special case for embeddings; if it's a tensor, it's only valid if we

299 # are considering size factors at constant scale, i.e., we just clone

300 # the tensor

tests/models/multimodal/processing/test_deepseek_ocr.py

+2 -4

@@ -108,8 +108,7 @@ class TestDeepseekOCREmptyImagesCrop:

108

109 def test_mismatched_image_size_raises(self, processor):

110 """Deliberately wrong image_size binding should still be caught

111 - by TensorSchema validation.

112 - """

111 + by TensorSchema validation."""

112 small_image = Image.new("RGB", (100, 100), color="green")

113

114 result = processor(

@@ -143,8 +142,7 @@ class TestDeepseekOCRInputValidation:

142 def test_mismatched_image_count_raises_value_error(self, processor):

143 """Two ``<image>`` tokens with one image must raise ValueError, not

144 AssertionError (which would yield HTTP 500 and, under ``-O``, silently

146 - drop the middle text segment).

147 - """

145 + drop the middle text segment)."""

146 image = Image.new("RGB", (100, 100), color="red")

147 with pytest.raises(ValueError, match="does not match number of images"):

148 processor(prompt="<image> and <image>", images=[image])

tests/models/multimodal/processing/test_llava_next_video.py

+2 -4

@@ -16,8 +16,7 @@ from ...utils import build_model_context

16 class _StubModel:

17 """Carries only the attributes the token-count method reads from

18 `self`, so the real methods can be exercised without constructing

19 - the full `nn.Module` (vision tower, language model, etc.).

20 - """

19 + the full `nn.Module` (vision tower, language model, etc.)."""

20

21 patch_grid_length: int

22 pooled_grid_length: int

@@ -32,8 +31,7 @@ get_mm_lora_token_counts = (

31 def test_num_mm_tokens_match_real_config(model_id):

32 """The stored grid lengths must match what `__init__` derives from

33 the real HF config, and the LoRA token counts must preserve the

35 - frame-level scaling exactly.

36 - """

34 + frame-level scaling exactly."""

35 ctx = build_model_context(model_id, limit_mm_per_prompt={"video": 1})

36 config = ctx.model_config.hf_config

37

tests/models/multimodal/processing/test_molmo2.py

+1 -2

@@ -70,8 +70,7 @@ def test_dummy_video_num_frames_override_honors_min_of_two(

70 ):

71 """A ``num_frames`` override below 2 must be ignored (the model needs at

72 least 2 frames), matching the "cannot be less than 2, will be ignored"

73 - warning.

74 - """

73 + warning."""

74 builder = object.__new__(Molmo2DummyInputsBuilder)

75 builder.info = SimpleNamespace(

76 get_hf_processor=lambda: SimpleNamespace(

tests/models/multimodal/processing/test_qwen3_vl.py

+4 -8

@@ -31,8 +31,7 @@ def _build_video_mm_data(

31

32 ``total_num_frames`` is set equal to the ndarray frame count so

33 that HF's ``sample_frames`` indices stay within bounds of the

34 - actual tensor that is passed.

35 - """

34 + actual tensor that is passed."""

35 video = np.zeros((num_frames, height, width, 3), dtype=np.uint8)

36 metadata = {

37 "fps": original_fps,

@@ -238,8 +237,7 @@ def _build_video_embeds_mm_data(

237 grid_thw: tuple[int, int, int] = (2, 4, 4),

238 ) -> dict[str, Any]:

239 """Create an embeds-only video item as an EPD consumer receives it:

241 - pre-computed embeddings plus the metadata published by the encoder.

242 - """

240 + pre-computed embeddings plus the metadata published by the encoder."""

241 import torch

242

243 t, h, w = grid_thw

@@ -258,8 +256,7 @@ def test_processor_video_embeds_with_timestamps(model_id: str) -> None:

256 """Embeds-only video input must size the placeholder range from the

257 grid and the real timestamps published by the encoder (EC consumer

258 path); synthesized or missing timestamps would change the token count

261 - and break embedding merging downstream.

262 - """

259 + and break embedding merging downstream."""

260 # `build_model_context` forces enable_mm_embeds off, so build the

261 # config directly; keep the same online-availability skip behavior.

262 HF_EXAMPLE_MODELS.find_hf_info(model_id).check_available_online(on_fail="skip")

@@ -304,8 +301,7 @@ def test_processor_video_embeds_with_timestamps(model_id: str) -> None:

301 def test_processor_video_embeds_missing_timestamps(model_id: str) -> None:

302 """Timestamps are required metadata for video embeds: they size the

303 placeholder range, so omitting them must fail loudly at parse time

307 - instead of silently producing a wrong prompt.

308 - """

304 + instead of silently producing a wrong prompt."""

305 ctx = build_model_context(

306 model_id,

307 limit_mm_per_prompt={"image": 0, "video": 1},

tests/models/multimodal/processing/test_transformers_audio.py

+1 -2

@@ -116,8 +116,7 @@ def test_audio_multimodal_processor(model_id, processor_cls):

116 @pytest.mark.parametrize("separator", [" and ", ""])

117 def test_audio_multiple_inputs(separator, processor_cls):

118 """Multiple audios per prompt are each detected as a separate placeholder

119 - and multi-modal item by the Transformers modelling backend.

120 - """

119 + and multi-modal item by the Transformers modelling backend."""

120 model_id = "ibm-granite/granite-speech-3.3-2b"

121 mm_processor = create_processor(model_id, processor_cls)

122

tests/models/multimodal/processing/test_transformers_image.py

+8 -16

@@ -83,8 +83,7 @@ def _process_two_images(processor_cls, separator: str):

83 @pytest.mark.parametrize("processor_cls", PROCESSOR_CLASSES)

84 def test_image_multiple_inputs(processor_cls):

85 """Multiple images per prompt are each detected as a separate placeholder

86 - and multi-modal item by the Transformers modelling backend.

87 - """

86 + and multi-modal item by the Transformers modelling backend."""

87 result = _process_two_images(processor_cls, separator="\n and ")

88

89 assert len(result["mm_placeholders"]["image"]) == 2

@@ -103,8 +102,7 @@ def test_image_adjacent_inputs(processor_cls):

102 @pytest.mark.parametrize("processor_cls", PROCESSOR_CLASSES)

103 def test_batch_padding_removed_from_image_items(processor_cls):

104 """Emu3 pads every image up to the largest in the batch, which would leave an

106 - item's data dependent on what it was processed with and so uncacheable.

107 - """

105 + item's data dependent on what it was processed with and so uncacheable."""

106 mm_processor = create_processor("BAAI/Emu3-Chat-hf", processor_cls)

107 image_token = mm_processor.info.get_hf_processor().image_token

108

@@ -146,8 +144,7 @@ def _process_one_gemma3_image(processor_cls):

144 @offsets_only

145 def test_non_embedding_tokens_excluded_from_placeholders():

146 """Gemma3 wraps each image in text that carries no embeddings, which must be

149 - inside the placeholder range but masked out of it.

150 - """

147 + inside the placeholder range but masked out of it."""

148 _, result = _process_one_gemma3_image(OffsetsMultiModalProcessor)

149

150 (placeholder,) = result["mm_placeholders"]["image"]

@@ -159,8 +156,7 @@ def test_legacy_placeholders_hold_only_image_tokens():

156 """The legacy path spans whatever `mm_token_type_ids` attributes to the image,

157 which for Gemma3 excludes the text wrapping it, unlike the replacement the offsets

158 path spans. Gemma3 is also the sharp case for the mask: its `image_token_id` is the

162 - marker in the unexpanded prompt, not the token the expansion repeats.

163 - """

159 + marker in the unexpanded prompt, not the token the expansion repeats."""

160 hf_processor, result = _process_one_gemma3_image(LegacyMultiModalProcessor)

161

162 (placeholder,) = result["mm_placeholders"]["image"]

@@ -176,8 +172,7 @@ def test_tokens_structuring_an_image_are_masked_not_dropped(processor_cls):

172 embeddings. Those belong inside the placeholder and masked out, because the token

173 count the processor reports is over the whole span. Idefics3 also refuses a prompt

174 holding `<image>` when no images are passed, which is how the offsets path has to

179 - tokenize it before splicing in the expansion.

180 - """

175 + tokenize it before splicing in the expansion."""

176 mm_processor = create_processor(

177 "HuggingFaceTB/SmolVLM-256M-Instruct", processor_cls

178 )

@@ -197,8 +192,7 @@ def test_tokens_structuring_an_image_are_masked_not_dropped(processor_cls):

192 @offsets_only

193 def test_missing_replacement_offsets_names_the_processor():

194 """A processor that reports no replacement offsets cannot be served, which must

200 - be said plainly rather than surfacing later as a field config mismatch.

201 - """

195 + be said plainly rather than surfacing later as a field config mismatch."""

196 model_id = "llava-hf/llava-onevision-qwen2-0.5b-ov-hf"

197 mm_processor = create_processor(model_id, OffsetsMultiModalProcessor)

198 hf_processor_cls = type(mm_processor.info.get_hf_processor())

@@ -282,8 +276,7 @@ def test_repeated_image_hits_the_processor_cache():

276 def test_spliced_prompt_matches_hf_expansion(model_id, prompt):

277 """The offsets path splices the expansion into a prompt tokenized without any

278 multi-modal data, so its token ids have to come out the same as the ones the HF

285 - processor produces itself, which is what the legacy path returns.

286 - """

279 + processor produces itself, which is what the legacy path returns."""

280 prompt_ids = []

281 for processor_cls in (LegacyMultiModalProcessor, OffsetsMultiModalProcessor):

282 mm_processor = create_processor(model_id, processor_cls)

@@ -305,8 +298,7 @@ def test_spliced_prompt_matches_hf_expansion(model_id, prompt):

298 def test_nested_image_fields_split_per_image(processor_cls):

299 """Idefics3 returns image fields with a leading batch dimension, putting the rows

300 belonging to each image one dimension further in. Slicing the batch dimension

308 - instead handed the first image every row and the second an empty tensor.

309 - """

301 + instead handed the first image every row and the second an empty tensor."""

302 mm_processor = create_processor(

303 "HuggingFaceTB/SmolVLM-256M-Instruct", processor_cls

304 )

tests/models/multimodal/processing/transformers_backend.py

+2 -4

@@ -29,8 +29,7 @@ PROCESSOR_CLASSES = [

29

30 def create_processor(model_id: str, processor_cls):

31 """Build a processor directly, because the registry only ever builds the one the

32 - installed transformers version selects, leaving the other path untested.

33 - """

32 + installed transformers version selects, leaving the other path untested."""

33 model_config = ModelConfig(model=model_id, model_impl="transformers")

34 ctx = InputProcessingContext(

35 model_config, cached_tokenizer_from_config(model_config)

@@ -41,8 +40,7 @@ def create_processor(model_id: str, processor_cls):

40

41 def create_cached_processor(model_id: str, processor_cls):

42 """Build a processor backed by a real multi-modal processor cache, and hand the

44 - cache back so a test can check it was actually used.

45 - """

43 + cache back so a test can check it was actually used."""

44 model_config = ModelConfig(model=model_id, model_impl="transformers")

45 model_config.multimodal_config.mm_processor_cache_gb = 4

46 ctx = InputProcessingContext(

tests/models/multimodal/test_cohere_asr.py

+1 -2

@@ -18,8 +18,7 @@ from vllm.model_executor.models.cohere_asr import CohereAsrForConditionalGenerat

18 def test_get_num_audio_tokens_streaming_estimate(audio_duration_s, expected_tokens):

19 """The duration-based estimate must convert ``window_stride`` (seconds) to a

20 sample hop and divide by the encoder subsampling factor. Values are pinned to

21 - concrete numbers so a rounding regression is actually caught.

22 - """

21 + concrete numbers so a rounding regression is actually caught."""

22 model_config = SimpleNamespace(

23 hf_config=SimpleNamespace(

24 preprocessor={"window_stride": 0.01, "sample_rate": 16000},

tests/models/multimodal/test_mapping.py

+1 -2

@@ -78,8 +78,7 @@ def test_cosmos3_modelopt_quantizer_weights_mapper():

78 """ModelOpt/Diffusers FP8 checkpoints ship native fake-quant buffers

79 (``*_quantizer._amax`` / ``._scale``) alongside the vLLM-consumable

80 ``weight_scale`` / ``input_scale`` sidecars. vLLM must drop the former

81 - (it has no parameter for them) while keeping the latter.

82 - """

81 + (it has no parameter for them) while keeping the latter."""

82 from vllm.model_executor.models.cosmos3 import Cosmos3ForConditionalGeneration

83

84 mapper = Cosmos3ForConditionalGeneration.hf_to_vllm_mapper

tests/models/multimodal/test_nano_nemotron_vl.py

+3 -6

@@ -81,8 +81,7 @@ class _FakeTensor:

81 """Sentinel stand-in for torch.Tensor in load_weights tests. Supports the

82 .detach().clone() chain used by load_weights for buffered mm weights;

83 both methods return self so identity (and the existing equality

84 - assertions) are preserved through cloning.

85 - """

84 + assertions) are preserved through cloning."""

85

86 def detach(self):

87 return self

@@ -166,8 +165,7 @@ def _make_mm_items_with_video_bytes(

165

166 def test_extract_audio_from_videos_passes_max_duration():

167 """_extract_audio_from_videos must forward VLLM_MAX_AUDIO_DECODE_DURATION_S

169 - to load_audio_pyav so decompression-bomb audio is rejected.

170 - """

168 + to load_audio_pyav so decompression-bomb audio is rejected."""

169 dummy_audio = (np.zeros(16000, dtype=np.float32), 16000.0)

170 mm_items = _make_mm_items_with_video_bytes(b"\x00" * 64)

171

@@ -187,8 +185,7 @@ def test_extract_audio_from_videos_passes_max_duration():

185

186 def test_extract_audio_from_videos_rejects_oversized_audio():

187 """When load_audio_pyav raises due to duration limit the video is

190 - marked as having no audio instead of crashing the server.

191 - """

188 + marked as having no audio instead of crashing the server."""

189 mm_items = _make_mm_items_with_video_bytes(b"\x00" * 64)

190

191 processor = object.__new__(NanoNemotronVLMultiModalProcessor)

tests/models/qwen4_exp/test_config.py

+1 -2

@@ -144,8 +144,7 @@ def test_qwen4_exp_mtp_override_sets_draft_config(

144 def test_qwen4_exp_rejects_pipeline_parallel_only_with_ple(ple_layer_ids) -> None:

145 """PLE needs raw input_ids, which non-first pipeline ranks never see. The

146 rest of the architecture is PP-capable, so the refusal must be conditional

147 - -- and must land before the engine spends time loading weights.

148 - """

147 + -- and must land before the engine spends time loading weights."""

148 vllm_config = SimpleNamespace(

149 model_config=SimpleNamespace(

150 hf_text_config=_text_config(ple_layer_ids=ple_layer_ids),

tests/models/test_deepseek_v4_fi_moe_ep.py

+5 -10

@@ -103,8 +103,7 @@ def test_fi_backend_strings_are_registered_mega_moe_backends():

103 @pytest.mark.parametrize("moe_backend", sorted(FLASHINFER_MOE_EP_BACKENDS))

104 def test_fi_moe_ep_backend_rejected_for_non_dsv4(moe_backend):

105 """An FI moe_ep backend with a non-DSv4 model must fail at config time

106 - instead of silently falling through to the generic FusedMoE path.

107 - """

106 + instead of silently falling through to the generic FusedMoE path."""

107 with pytest.raises(ValueError, match="only supported for DeepSeek-V4"):

108 validate_flashinfer_moe_ep_model(moe_backend, ["MixtralForCausalLM"])

109

@@ -120,8 +119,7 @@ def test_fi_moe_ep_backend_accepted_for_dsv4(moe_backend):

119 )

120 def test_native_deep_gemm_mega_moe_not_arch_gated(architectures):

121 """VLLM's own deep_gemm mega path is not DSv4-only (Kimi K3 uses it);

123 - models validate their own constraints at construction time.

124 - """

122 + models validate their own constraints at construction time."""

123 validate_flashinfer_moe_ep_model("deep_gemm_mega_moe", architectures)

124

125

@@ -134,8 +132,7 @@ def test_all_mega_backends_get_sequence_parallel_moe(moe_backend):

132 """Every mega backend must qualify for sequence-parallel MoE at

133 TP>1/EP: the predicate once matched only the native backend string,

134 which silently ran the fi backends full-batch with an all-reduce on

137 - every rank — 0.42-0.65x native e2e at TP8.

138 - """

135 + every rank — 0.42-0.65x native e2e at TP8."""

136 from vllm.models.deepseek_v4.nvidia.model import _use_sequence_parallel

137

138 vllm_config = SimpleNamespace(

@@ -179,8 +176,7 @@ def test_bootstrap_pins_the_device_vllm_bound(fake_flashinfer, monkeypatch):

176 """The runtime must not rederive the device from LOCAL_RANK/rank: under a

177 remapped CUDA_VISIBLE_DEVICES that ordinal points at the wrong GPU

178 (CUDA_ERROR_ILLEGAL_ADDRESS in the weight transforms). vLLM passes the

182 - device it already bound via BootstrapConfig.device.

183 - """

179 + device it already bound via BootstrapConfig.device."""

180 import vllm.utils.flashinfer_moe_ep as mod

181

182 pg = object()

@@ -245,8 +241,7 @@ def test_ckpt_uses_nvfp4_experts_reads_moe_quant_algo():

241

242 def test_dequant_fp4_ue8m0_gran32_decodes_lut_and_scales():

243 """One 32-element scale group per row: low nibble is the even element,

248 - high nibble the odd one, ue8m0 scale applies to the whole group.

249 - """

244 + high nibble the odd one, ue8m0 scale applies to the whole group."""

245 packed = torch.arange(32, dtype=torch.uint8).reshape(2, 16)

246 sf = torch.tensor([[127], [128]], dtype=torch.uint8) # 2**0, 2**1

247

tests/models/test_deepseek_v4_mega_moe.py

+2 -4

@@ -783,8 +783,7 @@ def test_deepseek_v4_mega_moe_stages_shared_scale_tma_layout(shared_block_m):

783 def test_deepseek_v4_pwal_hook_finalizes_mega_moe_and_mhc_broadcast():

784 """The loader invokes the model-level PWAL hook for every load format,

785 so it must finalize megamoe + mhc broadcast weights to cover dummy

786 - load, which skips load_weights().

787 - """

786 + load, which skips load_weights()."""

787 calls = []

788 stub = SimpleNamespace(

789 model=SimpleNamespace(

@@ -801,8 +800,7 @@ def test_deepseek_v4_pwal_hook_finalizes_mega_moe_and_mhc_broadcast():

800 def test_deepseek_v4_drafter_pwal_hooks_finalize_mega_moe():

801 """MTP/DSpark drafters load as their own top-level models, so each needs

802 its own PWAL hook now that the megamoe forward no longer finalizes

804 - weights lazily on first use.

805 - """

803 + weights lazily on first use."""

804 calls = []

805 mtp = SimpleNamespace(finalize_mega_moe_weights=lambda: calls.append("mtp"))

806 DeepSeekV4MTP.process_weights_after_loading(mtp)

tests/models/test_dspark_mla.py

+1 -2

@@ -215,8 +215,7 @@ def test_v41_dspark_loads_linear_scales(

215 ):

216 """Checkpoint ``.scale`` maps to the quant method's scale parameter and

217 loads untouched. MXFP8 block-scale expansion lives in the KMxfp8Static

218 - loader (see tests/quantization/test_modelopt.py), not in load_weights.

219 - """

218 + loader (see tests/quantization/test_modelopt.py), not in load_weights."""

219 from vllm.models.deepseek_v41.nvidia import dspark

220

221 mxfp8 = scale_dtype != torch.float32

tests/models/test_qwen3_5_mtp_config.py

+3 -6

@@ -73,8 +73,7 @@ def test_mtp_override_extracts_n_predict_from_multimodal_wrapper(

73 model_type: str, expected_arch: str

74 ) -> None:

75 """Verify that multimodal wrapper checkpoints with mtp_num_hidden_layers

76 - in text_config resolve n_predict and architecture correctly.

77 - """

76 + in text_config resolve n_predict and architecture correctly."""

77 cfg = SpeculativeConfig.hf_config_override(

78 _multimodal_wrapper_mtp_config(model_type, mtp_layers=2)

79 )

@@ -94,8 +93,7 @@ def test_mtp_override_top_level_precedence_over_nested_text_config(

93 model_type: str, expected_arch: str

94 ) -> None:

95 """Verify that an explicit top-level mtp_num_hidden_layers takes precedence

97 - over a nested text_config value.

98 - """

96 + over a nested text_config value."""

97 cfg = _multimodal_wrapper_mtp_config(model_type, mtp_layers=2)

98 cfg.mtp_num_hidden_layers = 3

99 overridden = SpeculativeConfig.hf_config_override(cfg)

@@ -115,8 +113,7 @@ def test_mtp_override_downloads_real_hf_hub_configs(

113 model_id: str, expected_arch: str

114 ) -> None:

115 """Verify that unmodified real-world checkpoints downloaded via

118 - AutoConfig.from_pretrained resolve n_predict=1 from text_config.

119 - """

116 + AutoConfig.from_pretrained resolve n_predict=1 from text_config."""

117 hf_config = AutoConfig.from_pretrained(model_id)

118 cfg = SpeculativeConfig.hf_config_override(hf_config)

119 assert cfg.model_type == "qwen3_5_mtp"

tests/models/test_utils.py

+1 -2

@@ -197,8 +197,7 @@ def test_merge_multimodal_embeddings_no_sync():

197 def test_get_rename_mapper_keeps_only_renames():

198 """`None` means "do not load", which is meaningless to the consumers of

199 this mapper (LoRA name parsing, quantization config layer lists), and

200 - applying it would silently shrink their lists.

201 - """

200 + applying it would silently shrink their lists."""

201 mapper = WeightsMapper(

202 orig_to_new_regex={re.compile(r"^drop_regex\."): None},

203 orig_to_new_substr={"drop_substr": None, "keep_substr": "kept"},

tests/models/test_vision.py

+1 -2

@@ -512,8 +512,7 @@ def _reference_input_norm(

512 def test_fused_input_norm_matches_reference(num_patches: int):

513 """FusedInputNorm must equal the plain affine, including for num_patches

514 above the cuDNN batch-norm grid limit (~65535) that previously raised

515 - CUDNN_STATUS_INTERNAL_ERROR (issue #51717).

516 - """

515 + CUDNN_STATUS_INTERNAL_ERROR (issue #51717)."""

516 channel = 3

517 patch_size = 14 * 14

518 image_mean = [0.48145466, 0.4578275, 0.40821073]

tests/models/transformers/fusers/test_linear.py

+23 -46

@@ -251,8 +251,7 @@ class UnderscoreInplaceParallelLinears(FourParallelLinears):

251

252 class PlainQKV(nn.Module):

253 """q/k/v returned as a bare tuple: no reshape stands between the calls and

254 - the return, so both QKVFuser and MergedColumnParallelFuser can rewrite it.

255 - """

254 + the return, so both QKVFuser and MergedColumnParallelFuser can rewrite it."""

255

256 def __init__(self, hidden: int = 32, heads: int = 4, head_dim: int = 8):

257 super().__init__()

@@ -339,8 +338,7 @@ class GuardedVAttention(FakeAttention):

338 """Gemma 4-style: `v_proj` is called once but also read in a `None` guard.

339

340 The guard is a fusion invariant (the fuser only binds instances where the

342 - projection exists), so it folds to `True` and the call still fuses.

343 - """

341 + projection exists), so it folds to `True` and the call still fuses."""

342

343 def forward(

344 self, hidden_states, attention_mask=None, past_key_values=None, **kwargs

@@ -370,8 +368,7 @@ class TruthyGuardAttention(FakeAttention):

368

369 An `nn.Linear` is always truthy, so the guard folds to `True` and the call

370 still fuses (the projection is a proven invariant, exactly as for a `None`

373 - guard).

374 - """

371 + guard)."""

372

373 def forward(

374 self, hidden_states, attention_mask=None, past_key_values=None, **kwargs

@@ -402,8 +399,7 @@ class StatementGuardVAttention(FakeAttention):

399 The `if`'s test is not an `ast.Compare`, so the identity-check scan skips it;

400 `_in_boolean_context` folds the test to `True` (an `nn.Linear` is always

401 truthy) and the call inside the now-`if True:` body still fuses, with the

405 - live `else` surviving the rewrite.

406 - """

402 + live `else` surviving the rewrite."""

403

404 def forward(

405 self, hidden_states, attention_mask=None, past_key_values=None, **kwargs

@@ -432,8 +428,7 @@ class BranchedKVAttention(FakeAttention):

428

429 The three calls do not share a block, so the fused GEMM must be hoisted to

430 the innermost block that dominates all of them (the function body). The

435 - live `if`/`else` must survive the rewrite.

436 - """

431 + live `if`/`else` must survive the rewrite."""

432

433 is_kv_shared_layer = False

434

@@ -464,8 +459,7 @@ class RebindArgAttention(FakeAttention):

459

460 The branch is dead at trace time so the match succeeds, but the source

461 rewrite must refuse: hoisting the GEMM above the rebind would feed k/v a

467 - different input. Result: no fusion.

468 - """

462 + different input. Result: no fusion."""

463

464 recompute = False

465

@@ -496,8 +490,7 @@ class InPlaceMutatedArgAttention(FakeAttention):

490 Like `RebindArgAttention`, but the mutation keeps the name in a `Load`

491 context (`hidden_states.mul_(2)`), so a bare `Name`-store check would miss it.

492 The rewrite must still refuse: hoisting the GEMM above the mutation changes

499 - the value k/v see. Result: no fusion.

500 - """

493 + the value k/v see. Result: no fusion."""

494

495 recompute = False

496

@@ -527,8 +520,7 @@ class AliasMutatedArgAttention(FakeAttention):

520

521 `hidden_states.view(-1)` shares storage with its base, so `alias.mul_(2)`

522 changes what k/v read without ever naming `hidden_states` as a target. Only

530 - tracking names that may alias the input catches it. Result: no fusion.

531 - """

523 + tracking names that may alias the input catches it. Result: no fusion."""

524

525 recompute = False

526

@@ -559,8 +551,7 @@ class FunctionalInplaceArgAttention(FakeAttention):

551

552 `F.relu(hidden_states, inplace=True)` writes through an argument, so the

553 name appears only in a `Load` context inside a call that is not a method on

562 - it -- invisible to a trailing-underscore method check. Result: no fusion.

563 - """

554 + it -- invisible to a trailing-underscore method check. Result: no fusion."""

555

556 recompute = False

557

@@ -590,8 +581,7 @@ class SubscriptMutatedArgAttention(FakeAttention):

581

582 `hidden_states[..., 0] = 0` leaves the name in a `Load` context on the

583 subscript's value, so it too evades a bare `Name`-store check. The rewrite

593 - must refuse for the same reason. Result: no fusion.

594 - """

584 + must refuse for the same reason. Result: no fusion."""

585

586 recompute = False

587

@@ -621,8 +611,7 @@ class NonFoldableRefAttention(FakeAttention):

611

612 Deleting `v_proj` would break `isinstance(self.v_proj, nn.Linear)`: it reads

613 the projection's type, not just whether it exists, so the fuser cannot fold

624 - it and must refuse rather than rewrite it.

625 - """

614 + it and must refuse rather than rewrite it."""

615

616 def forward(

617 self, hidden_states, attention_mask=None, past_key_values=None, **kwargs

@@ -847,8 +836,7 @@ def test_glu_identifies_down_projection():

836

837 It is forced to `RowParallelLinear` in `update_attrs` so its sharded input

838 matches the column-parallel merged gate/up; `None` when there is no such

850 - projection to force (fusion of gate/up still applies).

851 - """

839 + projection to force (fusion of gate/up still applies)."""

840 with torch.device("meta"):

841 assert get_fuser(GLUMLP(), GLUFuser).down_name == "down_proj"

842 assert get_fuser(ReversedGLUMLP(), GLUFuser).down_name == "down_proj"

@@ -1009,8 +997,7 @@ def test_bypass_existence_guard_folds_identity_none_check(expr, expected):

997 The projection is a proven invariant (it exists as an `nn.Linear`), so an

998 `is (not) None` guard has a constant truth value; folding it lets the call be

999 rewritten away without changing the guard's outcome. `==`/`!=` are excluded

1012 - (see the refusal test) because a subclass may override `__eq__`/`__ne__`.

1013 - """

1000 + (see the refusal test) because a subclass may override `__eq__`/`__ne__`."""

1001 funcdef, ref = _guard_funcdef(expr)

1002 bypass_existence_guard(funcdef, ref, "v_proj")

1003 test = next(node for node in ast.walk(funcdef) if isinstance(node, ast.IfExp)).test

@@ -1029,8 +1016,7 @@ def test_bypass_existence_guard_folds_bare_truthiness(expr):

1016

1017 An `nn.Linear` has no `__bool__`/`__len__`, so it is always truthy; in a pure

1018 boolean position (a test, or a `not` operand) the reference can be replaced by

1032 - `True`, letting the projection be deleted.

1033 - """

1019 + `True`, letting the projection be deleted."""

1020 funcdef, ref = _guard_funcdef(expr)

1021 bypass_existence_guard(funcdef, ref, "v_proj")

1022 # No `self.v_proj` reference survives; a literal `True` took its place.

@@ -1045,8 +1031,7 @@ def test_bypass_existence_guard_folds_bare_if_statement():

1031

1032 The `len(node.ops) == 1` `Compare` scan skips it (it is not a comparison at

1033 all); `_in_boolean_context` then matches the `if`'s test and folds it, so the

1048 - bare-truthiness statement form is handled, not silently dropped.

1049 - """

1034 + bare-truthiness statement form is handled, not silently dropped."""

1035 funcdef = ast.parse(

1036 "def f(self, a, b):\n if self.v_proj:\n return a\n return b"

1037 ).body[0]

@@ -1078,8 +1063,7 @@ def test_bypass_existence_guard_refuses_non_guard_reference(expr):

1063 Rewriting it is unsafe (its value depends on more than the projection's

1064 existence), so the fuser refuses rather than change semantics. `==`/`!=` are

1065 refused too: `is_linear` accepts `nn.Linear` subclasses that could override

1081 - `__eq__`/`__ne__`, so equality need not agree with `is (not) None`.

1082 - """

1066 + `__eq__`/`__ne__`, so equality need not agree with `is (not) None`."""

1067 funcdef, ref = _guard_funcdef(expr)

1068 with pytest.raises(ValueError, match="outside an existence guard"):

1069 bypass_existence_guard(funcdef, ref, "v_proj")

@@ -1119,8 +1103,7 @@ def test_merged_column_fuser_folds_existence_guard():

1103

1104 Guard folding lives on `StackedFuser`, not just `QKVFuser`, because every

1105 stacked fuser deletes the projections it merges -- a surviving reference to

1122 - one would fail at runtime.

1123 - """

1106 + one would fail at runtime."""

1107 with torch.device("meta"):

1108 module = GuardedParallelLinears()

1109 fuser = MergedColumnParallelFuser.match(trace(module), module)

@@ -1146,8 +1129,7 @@ def test_merged_column_fuser_rejects_input_mutation(cls):

1129 The mutation may be direct (`x.add_(1)`), through a view that shares the

1130 input's storage, or through a call that writes an argument in place

1131 (`F.relu(x, inplace=True)`, `torch.relu_(x)`). Each one silently changes the

1149 - fused result, so the rewrite must refuse rather than fuse.

1150 - """

1132 + fused result, so the rewrite must refuse rather than fuse."""

1133 with torch.device("meta"):

1134 module = cls()

1135 fuser = MergedColumnParallelFuser.match(trace(module), module)

@@ -1205,8 +1187,7 @@ def test_merged_column_fuser_supports_any_number_of_linears(

1187

1188 def test_merged_name_is_unique_per_linear_names():

1189 """The computed name must not collide across different projections, so

1208 - stacking two unrelated fusions can never clobber packed_modules_mapping.

1209 - """

1190 + stacking two unrelated fusions can never clobber packed_modules_mapping."""

1191 ab_c = MergedColumnParallelFuser(source_cls="M", linear_names=("a_b", "c"))

1192 a_bc = MergedColumnParallelFuser(source_cls="M", linear_names=("a", "b_c"))

1193 assert ab_c.merged_name != a_bc.merged_name

@@ -1222,8 +1203,7 @@ def test_heterogeneous_instances_fall_back_to_merged_column_fuser(

1203 layer whose out_features aren't a multiple of its own head_dim (as a

1204 misconfigured or heterogeneous checkpoint might have) fails QKVFuser's

1205 validation and must still fuse via the generic fallback, while a

1225 - well-formed sibling of the same class keeps using QKVFuser.

1226 - """

1206 + well-formed sibling of the same class keeps using QKVFuser."""

1207 with torch.device("meta"):

1208 container = nn.Module()

1209 container.compatible = PlainQKV(head_dim=8)

@@ -1244,8 +1224,7 @@ def test_detects_and_rewrites_packed_qkv(kv_heads):

1224 """A single projection split into q/k/v must be re-sharded, not merged.

1225

1226 Only the split sizes change: `QKVParallelLinear` loads the packed

1247 - checkpoint weight as-is, and shards q by heads while replicating k/v.

1248 - """

1227 + checkpoint weight as-is, and shards q by heads while replicating k/v."""

1228 with torch.device("meta"):

1229 meta = PackedQKVAttention(kv_heads=kv_heads)

1230 fuser = get_fuser(meta, PackedQKVFuser)

@@ -1276,8 +1255,7 @@ def test_detects_and_rewrites_packed_qkv(kv_heads):

1255

1256 def test_per_head_split_is_not_packed_qkv():

1257 """The split must consume the whole projection, else its sizes are head

1279 - widths and re-sharding by them would be wrong.

1280 - """

1258 + widths and re-sharding by them would be wrong."""

1259 with torch.device("meta"):

1260 assert get_fuser(PerHeadSplitAttention(), PackedQKVFuser) is None

1261

@@ -1397,8 +1375,7 @@ def test_act_and_mul_derived_from_module(default_vllm_config):

1375

1376 def _wider_model_config(head_dim: int) -> SimpleNamespace:

1377 """A model whose global head size is twice `head_dim`, as a wider layer

1400 - elsewhere in a heterogeneous checkpoint would make it.

1401 - """

1378 + elsewhere in a heterogeneous checkpoint would make it."""

1379 return SimpleNamespace(

1380 model_config=SimpleNamespace(get_head_size=lambda: 2 * head_dim),

1381 quant_config=None,

tests/models/transformers/fusers/test_mla.py

+5 -10

@@ -45,8 +45,7 @@ def _attention(q_lora_rank: int | None):

45

46 def _match(q_lora_rank: int | None) -> MLAFuser | None:

47 """Match a meta `DeepseekV2Attention` directly (bypassing the per-class

48 - `get_fuser` cache, so both q_lora variants of the same class are seen).

49 - """

48 + `get_fuser` cache, so both q_lora variants of the same class are seen)."""

49 attn = _attention(q_lora_rank)

50 return MLAFuser.match(trace(attn), attn)

51

@@ -82,8 +81,7 @@ def test_discovers_modules_with_q_lora():

81

82 def test_q_lora_stacks_qkv_a_proj():

83 """The MLA layer reads `q_a_proj` and `kv_a_proj_with_mqa` fused into one

85 - down-projection, so both checkpoint weights must remap into it.

86 - """

84 + down-projection, so both checkpoint weights must remap into it."""

85 fuser = _match(q_lora_rank=64)

86 assert isinstance(fuser, MLAFuser)

87 prefix = "model.layers.0.self_attn"

@@ -104,8 +102,7 @@ def test_update_forward_rewrites_real_attention(q_lora_rank):

102 The q-LoRA path hoists the merged down-projection to a top-level statement

103 while its calls sit inside the `q_lora_rank` branches, so the region it

104 spans is wide; a check that over-approximates aliasing would silently refuse

107 - here and cost the fusion rather than fail loudly.

108 - """

105 + here and cost the fusion rather than fail loudly."""

106 fuser = _match(q_lora_rank)

107 assert isinstance(fuser, MLAFuser)

108 fuser.update_forward(_attention(q_lora_rank))

@@ -129,8 +126,7 @@ class _Norm(nn.Module):

126

127 class RenamedMLA(nn.Module):

128 """An MLA-shaped attention whose children have non-standard names, proving

132 - discovery is by structure and not attribute name.

133 - """

129 + discovery is by structure and not attribute name."""

130

131 def __init__(self):

132 super().__init__()

@@ -157,8 +153,7 @@ class RenamedMLA(nn.Module):

153

154 def test_discovers_modules_under_arbitrary_names():

155 """Discovery is purely structural: `RenamedMLA` gives its children non-standard

160 - names, and `match` still locates each projection by dataflow.

161 - """

156 + names, and `match` still locates each projection by dataflow."""

157 with torch.device("meta"):

158 module = RenamedMLA()

159 fuser = MLAFuser.match(trace(module), module)

tests/models/transformers/fusers/test_moe.py

+6 -12

@@ -277,8 +277,7 @@ def test_moe_fuser_matches_scaled_router():

277

278 def test_moe_fuser_matches_grouped_router():

279 """Group masking between the score and the routing top-k still matches: the

280 - scorer is anchored on the routing (last) top-k, not the group one.

281 - """

280 + scorer is anchored on the routing (last) top-k, not the group one."""

281 with torch.device("meta"):

282 block = MoEBlock(GroupedRouter)

283 fuser = MoEBlockFuser.match(block, "experts")

@@ -288,8 +287,7 @@ def test_moe_fuser_matches_grouped_router():

287

288 def test_moe_fuser_matches_correction_router():

289 """A score-correction bias buffer (DeepSeek-V3 noaux) is allowed; the

291 - rebuilt gate carries it in fp32 for FusedMoE's biased routers.

292 - """

290 + rebuilt gate carries it in fp32 for FusedMoE's biased routers."""

291 with torch.device("meta"):

292 block = MoEBlock(CorrectionRouter)

293 fuser = MoEBlockFuser.match(block, "experts")

@@ -300,8 +298,7 @@ def test_moe_fuser_matches_correction_router():

298 def test_moe_fuser_reads_router_dtype_from_the_gate():

299 """A router that computes its logits in fp32 must keep routing in fp32 when

300 rebuilt, even though no config field names the dtype. The cast back to the

303 - activation dtype after the top-k must not be mistaken for the routing dtype.

304 - """

301 + activation dtype after the top-k must not be mistaken for the routing dtype."""

302 with torch.device("meta"):

303 fp32 = MoEBlockFuser.match(MoEBlock(Fp32Router), "experts")

304 default = MoEBlockFuser.match(MoEBlock(TopKRouter), "experts")

@@ -320,8 +317,7 @@ def test_moe_fuser_detects_shared_experts():

317

318 def test_moe_fuser_skips_shared_detection_without_extra_children():

319 """With only experts and gate, shared-expert detection (and its block trace)

323 - is skipped, so a gate-derived add is not misread as a shared expert.

324 - """

320 + is skipped, so a gate-derived add is not misread as a shared expert."""

321 with torch.device("meta"):

322 block = MoEBlockNoShared()

323 fuser = MoEBlockFuser.match(block, "experts")

@@ -366,8 +362,7 @@ def test_moe_fuser_declines_unsupported(block_cls):

362

363 def test_moe_fuser_ignores_nested_returns():

364 """A tuple `return` inside a nested helper must not decline a block whose own

369 - forward returns a single tensor.

370 - """

365 + forward returns a single tensor."""

366 with torch.device("meta"):

367 block = MoEBlockNestedTupleReturn()

368 assert isinstance(MoEBlockFuser.match(block, "experts"), MoEBlockFuser)

@@ -375,8 +370,7 @@ def test_moe_fuser_ignores_nested_returns():

370

371 def test_moe_fuser_router_requires_connected_dataflow():

372 """A gate with linear + softmax + top-k present but not wired as a router

378 - (top-k selects over the input, not the scored logits) is not detected.

379 - """

373 + (top-k selects over the input, not the scored logits) is not detected."""

374 with torch.device("meta"):

375 block = MoEBlock(DisconnectedRouter)

376 assert MoEBlockFuser.match(block, "experts") is None

tests/models/transformers/fusers/test_rms_norm.py

+9 -18

@@ -83,8 +83,7 @@ class ToleranceRMSNorm(NamedEpsRMSNorm):

83

84 class LiteralEpsRMSNorm(RMSNorm):

85 """eps is a literal in the source, and an unrelated attribute happens to

86 - hold the same value: matching on the value alone would bind to it.

87 - """

86 + hold the same value: matching on the value alone would bind to it."""

87

88 def __init__(self, hidden: int = 16, eps: float = 1e-5):

89 super().__init__(hidden, eps)

@@ -96,8 +95,7 @@ class LiteralEpsRMSNorm(RMSNorm):

95

96 class AmbiguousEpsRMSNorm(nn.Module):

97 """Two attributes hold the eps value, and the forward reads the second, so

99 - their order cannot pick it.

100 - """

98 + their order cannot pick it."""

99

100 def __init__(self, hidden: int = 16, eps: float = 1e-6):

101 super().__init__()

@@ -233,8 +231,7 @@ def test_weightless_norm_has_no_hidden_size(default_vllm_config):

231

232 def test_fused_rms_norm_op_default_eps(default_vllm_config):

233 """`torch.nn.RMSNorm` (a single `F.rms_norm` call) matches via the fast path;

236 - its default `eps=None` resolves to `finfo(dtype).eps` in `fuse`.

237 - """

234 + its default `eps=None` resolves to `finfo(dtype).eps` in `fuse`."""

235 from vllm.model_executor.layers.layernorm import RMSNorm as VLLMRMSNorm

236

237 with torch.device("meta"):

@@ -250,8 +247,7 @@ def test_fused_rms_norm_op_default_eps(default_vllm_config):

247

248 def test_eps_is_derived_per_instance(default_vllm_config):

249 """Two instances of the same norm class with different eps must fuse to their

253 - own eps: the type-cached fuser holds only structure, not this value.

254 - """

250 + own eps: the type-cached fuser holds only structure, not this value."""

251 with torch.device("meta"):

252 for eps in (1e-5, 1e-6):

253 module = RMSNorm(16, eps=eps)

@@ -263,8 +259,7 @@ def test_eps_is_derived_per_instance(default_vllm_config):

259 @pytest.mark.parametrize("cls", [NamedEpsRMSNorm, ToleranceRMSNorm])

260 def test_eps_attr_is_found_by_value_not_name(cls, default_vllm_config):

261 """The eps attribute is identified by holding the traced value, so a norm

266 - stays per-instance correct whatever it names it.

267 - """

262 + stays per-instance correct whatever it names it."""

263 with torch.device("meta"):

264 for eps in (1e-5, 1e-6):

265 module = cls(16, eps=eps)

@@ -276,8 +271,7 @@ def test_eps_attr_is_found_by_value_not_name(cls, default_vllm_config):

271

272 def test_literal_eps_is_not_mistaken_for_an_attribute(default_vllm_config, caplog):

273 """A literal eps is recognised as coming from no attribute, even when one

279 - holds the same value, and is taken from the traced source instead.

280 - """

274 + holds the same value, and is taken from the traced source instead."""

275 logger = "vllm.model_executor.models.transformers.fusers.rms_norm"

276 with caplog.at_level("DEBUG", logger=logger), torch.device("meta"):

277 module = LiteralEpsRMSNorm()

@@ -290,8 +284,7 @@ def test_literal_eps_is_not_mistaken_for_an_attribute(default_vllm_config, caplo

284

285 def test_ambiguous_eps_attrs_are_disambiguated(default_vllm_config):

286 """When several attributes hold the eps value, the one the forward actually

293 - reads is identified, and they are all left as they were found.

294 - """

287 + reads is identified, and they are all left as they were found."""

288 with torch.device("meta"):

289 module = AmbiguousEpsRMSNorm(16, eps=1e-6)

290 before = dict(vars(module))

@@ -308,8 +301,7 @@ def test_ambiguous_eps_attrs_are_disambiguated(default_vllm_config):

301 def test_fused_norm_is_gather_capable(default_vllm_config):

302 """Every weighted fused norm is emitted gather-capable, so a norm on a head-sharded

303 projection (OLMoE-style) self-corrects at runtime with no QKV-specific

311 - plumbing. A full-width input skips the gather and equals a plain norm.

312 - """

304 + plumbing. A full-width input skips the gather and equals a plain norm."""

305 from vllm.model_executor.layers.layernorm import GemmaRMSNorm, RMSNorm

306 from vllm.model_executor.models.transformers.fusers import rms_norm

307

@@ -331,8 +323,7 @@ def test_fused_norm_is_gather_capable(default_vllm_config):

323

324 def test_gathered_norm_rejects_uneven_sharding(default_vllm_config):

325 """A sharded input (narrower than the full-width weight) that does not tile

334 - the weight evenly across ranks is rejected before any collective.

335 - """

326 + the weight evenly across ranks is rejected before any collective."""

327 from vllm.model_executor.models.transformers.fusers import rms_norm

328

329 norm = rms_norm.TPAwareRMSNorm(hidden_size=8, eps=1e-6)

tests/models/transformers/test_layer_registry.py

+2 -4

@@ -28,8 +28,7 @@ def fake_hw_layernorm(monkeypatch):

28 A `SimpleNamespace` stands in for the module: `importlib.import_module`

29 returns it from `sys.modules` and `getattr` resolves `RMSNorm`, while its

30 attributes are set at construction (no `ModuleType` attribute-set that mypy

31 - rejects, no constant `setattr` that ruff rejects).

32 - """

31 + rejects, no constant `setattr` that ruff rejects)."""

32 module = types.SimpleNamespace(RMSNorm=type("HwRMSNorm", (), {}))

33 monkeypatch.setitem(sys.modules, HW_MODULE, module)

34 return module

@@ -82,8 +81,7 @@ def test_act_and_mul_falls_back_for_unknown_activation(

81 @pytest.fixture(scope="module")

82 def tiny_llama_path(tmp_path_factory):

83 """A randomly-initialized microscopic Llama saved to disk (with an ungated

85 - tokenizer) so vLLM can load it like any local checkpoint.

86 - """

84 + tokenizer) so vLLM can load it like any local checkpoint."""

85 from transformers import AutoTokenizer, LlamaConfig, LlamaForCausalLM

86

87 tokenizer = AutoTokenizer.from_pretrained("hf-internal-testing/llama-tokenizer")

tests/multimodal/media/test_audio.py

+6 -12

@@ -54,8 +54,7 @@ def test_audio_media_io_load_base64(dummy_audio_bytes):

54 def test_audio_media_io_load_base64_rejects_malformed(dummy_audio_bytes):

55 """Malformed base64 must surface as a ValueError so the server answers 400.

56 Without strict decoding the bad characters are dropped, the garbage reaches

57 - libsndfile, and the client gets a 500 instead.

58 - """

57 + libsndfile, and the client gets a 500 instead."""

58 encoded = base64.b64encode(dummy_audio_bytes).decode("utf-8")

59 malformed = encoded[:8] + "!!!@@@###" + encoded[8:]

60

@@ -134,8 +133,7 @@ def test_small_file_passes_memory_guard():

133

134 def test_memory_guard_rejects_large_allocation():

135 """A file whose frames*channels*4 exceeds the byte limit must be

137 - rejected before allocating the buffer.

138 - """

136 + rejected before allocating the buffer."""

137 # 100_000 frames * 8 channels * 4 bytes = 3.2 MB

138 payload = _make_flac_bytes(frames=100_000, channels=8, samplerate=48000)

139 # Set limit to 1 MiB — should reject

@@ -150,8 +148,7 @@ def test_memory_guard_rejects_large_allocation():

148

149 def test_forged_samplerate_rejected_by_memory_guard():

150 """The PoC scenario: high sample rate fools the duration guard but

153 - the memory guard catches the large frame*channel allocation.

154 - """

151 + the memory guard catches the large frame*channel allocation."""

152 # Forged high sample rate: 655350 Hz, 8 channels, 1M frames

153 # Duration guard sees: 1_000_000 / 655_350 = 1.5s → passes

154 # Memory: 1_000_000 * 8 * 4 = 32 MB

@@ -212,8 +209,7 @@ def test_load_audio_unknown_backend_rejected(dummy_audio_bytes):

209

210 def test_load_audio_auto_falls_back_without_torchcodec(dummy_audio_bytes):

211 """`auto` must fall back to the soundfile → PyAV chain when torchcodec

215 - is not importable.

216 - """

212 + is not importable."""

213 ref_audio, ref_sr = load_audio_soundfile(BytesIO(dummy_audio_bytes), sr=None)

214 with patch.object(audio_module, "load_audio_torchcodec", side_effect=ImportError):

215 audio, sr = load_audio(BytesIO(dummy_audio_bytes), sr=None, backend="auto")

@@ -223,8 +219,7 @@ def test_load_audio_auto_falls_back_without_torchcodec(dummy_audio_bytes):

219

220 def test_load_audio_auto_falls_back_without_ffmpeg(dummy_audio_bytes):

221 """Torchcodec installed but system ffmpeg missing (`AudioDecoder is None`)

226 - must surface as ImportError so `auto` falls back to soundfile → PyAV.

227 - """

222 + must surface as ImportError so `auto` falls back to soundfile → PyAV."""

223 ref_audio, ref_sr = load_audio_soundfile(BytesIO(dummy_audio_bytes), sr=None)

224 with patch.object(audio_module, "AudioDecoder", None):

225 with pytest.raises(ImportError, match="torchcodec audio backend"):

@@ -239,8 +234,7 @@ def test_load_audio_auto_falls_back_when_libtorchcodec_unloadable(

234 ):

235 """Torchcodec loads its ffmpeg-backed core lazily at decoder construction;

236 a "Could not load libtorchcodec" RuntimeError there (no system ffmpeg)

242 - must also surface as ImportError so `auto` falls back.

243 - """

237 + must also surface as ImportError so `auto` falls back."""

238

239 class _UnloadableDecoder:

240 def __init__(self, *args, **kwargs):

tests/multimodal/media/test_connector.py

+1 -2

@@ -228,8 +228,7 @@ async def test_fetch_image_local_files_with_space_in_name(image_url: str):

228 @pytest.mark.asyncio

229 async def test_fetch_image_data_url_with_params():

230 """RFC 2397 allows parameters between the mediatype and the base64

231 - marker; they must not be rejected or leak into the media type.

232 - """

231 + marker; they must not be rejected or leak into the media type."""

232 connector = MediaConnector()

233

234 image = Image.new("RGB", (4, 4), color=(255, 0, 0))

tests/multimodal/media/test_video.py

+5 -10

@@ -398,8 +398,7 @@ def test_pynvvideocodec_unrelated_error_propagates(

398

399 class TestMergeKwargsGpuBackendPolicy:

400 """Verify that merge_kwargs blocks request-level GPU backend selection

401 - when the static (engine-level) config did not configure that backend.

402 - """

401 + when the static (engine-level) config did not configure that backend."""

402

403 def test_pynvvideocodec_requires_gpu(self):

404 assert VIDEO_LOADER_REGISTRY.backend_requires_gpu(PYNVVIDEOCODEC_VIDEO_BACKEND)

@@ -492,8 +491,7 @@ class TestMergeKwargsGpuBackendPolicy:

491

492 def test_static_pynv_with_different_runtime_gpu_backend(self):

493 """If static sets pynv via video_backend but runtime tries to set it

495 - via the codec-level 'backend' key (without a static match), strip it.

496 - """

494 + via the codec-level 'backend' key (without a static match), strip it."""

495 result = VideoMediaIO.merge_kwargs(

496 default_kwargs={"video_backend": "pynvvideocodec"},

497 runtime_kwargs={"backend": "pynvvideocodec"},

@@ -532,8 +530,7 @@ class TestMergeKwargsGpuBackendPolicy:

530 def test_strips_request_level_device(self):

531 """The decode device is a startup-only knob: a request must not move

532 decoding onto the GPU when the startup config did not opt in, nor off

535 - it when it did.

536 - """

533 + it when it did."""

534 result = VideoMediaIO.merge_kwargs(

535 default_kwargs={"backend": "torchcodec"},

536 runtime_kwargs={"device": "cuda"},

@@ -558,8 +555,7 @@ class TestMergeKwargsGpuBackendPolicy:

555 def test_switching_backend_drops_stale_codec_options(default_kwargs):

556 """Codec-specific options from the static config must not leak into a

557 different codec backend selected per-request, where they would fail the

561 - new backend's option validation.

562 - """

558 + new backend's option validation."""

559 result = VideoMediaIO.merge_kwargs(

560 default_kwargs={**default_kwargs, "num_frames": 8},

561 runtime_kwargs={"backend": "opencv"},

@@ -581,8 +577,7 @@ def test_pynvvc_frames_normalized_to_nhwc(layout: str):

577 """PyNvVideoCodec frame batches are normalized to NHWC regardless of the

578 per-frame layout the decoder emits (it has varied across versions), so the

579 HF video processors (which materialize a PIL image per frame) receive the

584 - same NHWC shape as every other video backend.

585 - """

580 + same NHWC shape as every other video backend."""

581 torch = pytest.importorskip("torch")

582

583 n, h, w, c = 4, 5, 6, 3

tests/multimodal/test_hasher.py

+1 -2

@@ -141,8 +141,7 @@ def test_hash_collision_video_num_frames():

141

142 def test_hash_video_tensor_frames():

143 """Videos holding tensor frames (e.g. NVDEC-decoded) hash like

144 - array-framed ones, from the original bytes without a D2H copy.

145 - """

144 + array-framed ones, from the original bytes without a D2H copy."""

145 source = b"x" * 100

146

147 def item_for_hash(frames):

tests/multimodal/test_image.py

+1 -2

@@ -80,8 +80,7 @@ def test_l_mode_no_trns_to_rgb():

80

81 def test_exif_transpose_normalizes_orientation():

82 """Image with EXIF orientation 3 (180-degree rotation) should be

83 - normalized so pixel data matches visual display.

84 - """

83 + normalized so pixel data matches visual display."""

84 # Built synthetically because the existing assets are PNGs which

85 # don't carry EXIF orientation metadata; we need a JPEG with an

86 # injected EXIF orientation=3 tag.

tests/multimodal/test_parse.py

+2 -4

@@ -59,8 +59,7 @@ def test_frame_size_hwc_chw(frame):

59 def test_video_with_metadata_tensor_passthrough():

60 """Tensor frames pass through unchanged regardless of device: HF video

61 processors accept tensors, and device-resident frames (e.g. NVDEC-decoded)

62 - must not be copied back to host.

63 - """

62 + must not be copied back to host."""

63 frames = torch.zeros((4, H, W, 3), dtype=torch.uint8)

64 video, metadata = MultiModalDataParser()._get_video_with_metadata(frames)

65

@@ -71,8 +70,7 @@ def test_video_with_metadata_tensor_passthrough():

70 @pytest.mark.skipif(not torch.cuda.is_available(), reason="Requires CUDA")

71 def test_video_with_metadata_keeps_device_tensor():

72 """Device-resident frames (e.g. NVDEC-decoded) pass through as tensors,

74 - so a device-side HF processor can consume them without a D2H copy.

75 - """

73 + so a device-side HF processor can consume them without a D2H copy."""

74 frames = torch.zeros((4, H, W, 3), dtype=torch.uint8, device="cuda")

75 video, metadata = MultiModalDataParser()._get_video_with_metadata(frames)

76

tests/multimodal/test_registry.py

+1 -2

@@ -57,8 +57,7 @@ def test_dummy_inputs_scheduler_budget(

57 @pytest.mark.core_model

58 def test_supports_multimodal_inputs(model_id, limit_mm_per_prompt, expected):

59 """Test supports_multimodal_inputs returns correct boolean for various

60 - configs.

61 - """

60 + configs."""

61 ctx = build_model_context(

62 model_id,

63 limit_mm_per_prompt=limit_mm_per_prompt,

tests/multimodal/test_utils.py

+1 -2

@@ -231,8 +231,7 @@ def test_group_and_batch_mm_items_splits_shared_data_by_dtype():

231

232 def test_encode_image_url_uncommon_format_has_valid_mimetype():

233 """A format missing from mimetypes.types_map must still produce a

234 - well-formed type/subtype media type, not a bare 'image'.

235 - """

234 + well-formed type/subtype media type, not a bare 'image'."""

235 # ".tga" is not registered in mimetypes.types_map, so it exercises the

236 # fallback path; Pillow can still encode it.

237 url = encode_image_url(Image.new("RGB", (2, 2)), format="TGA")

tests/multimodal/test_video.py

+10 -20

@@ -1214,8 +1214,7 @@ def test_torchcodec_backend_rejects_frame_recovery(dummy_video_path):

1214 @pytest.mark.skipif(not current_platform.is_cuda(), reason="Requires CUDA")

1215 def test_torchcodec_backend_cuda_decodes_on_gpu(dummy_video_path):

1216 """With device="cuda", torchcodec decodes via NVDEC and keeps the frames

1217 - on the GPU instead of returning a host-side numpy array.

1218 - """

1217 + on the GPU instead of returning a host-side numpy array."""

1218 pytest.importorskip("torchcodec")

1219

1220 with open(dummy_video_path, "rb") as f:

@@ -1508,8 +1507,7 @@ def test_glm46v_duration_estimation_from_fps():

1507 class TestGLMGASamplingCaps:

1508 """Regression tests for GHSA-58v5-2m8f-94pr: request-controlled fps

1509 and max_frames must not create intermediate allocations larger than

1511 - the actual frame count.

1512 - """

1510 + the actual frame count."""

1511

1512 @staticmethod

1513 def _source(

@@ -1563,8 +1561,7 @@ def test_glm5next_backend_selected_for_processor():

1561 """Glm5NextVideoProcessor maps to the glm5next loader so only the

1562 sampled frames are decoded instead of the whole container. Both the

1563 borrowed-config spelling and the dedicated Glm5next class name (landing

1566 - with the new checkpoint) must resolve.

1567 - """

1564 + with the new checkpoint) must resolve."""

1565 for name in ("Glm5NextVideoProcessor", "Glm5nextVideoProcessor"):

1566 assert VIDEO_LOADER_REGISTRY.get_backend_for_video_processor(name) == "glm5next"

1567

@@ -1586,8 +1583,7 @@ def test_glm5next_backend_indices_match_sampler(

1583 total_frames, original_fps, duration, fps, max_frames

1584 ):

1585 """The loader must select exactly the frames the processor's sampler

1589 - would, with target.fps mapping onto the raw-fps override.

1590 - """

1586 + would, with target.fps mapping onto the raw-fps override."""

1587 from vllm.transformers_utils.processors.glm5next import (

1588 glm_sample_frame_indices,

1589 )

@@ -1615,8 +1611,7 @@ def test_glm5next_backend_indices_match_sampler(

1611

1612 def test_glm5next_backend_metadata_contract():

1613 """create_hf_metadata reports the subset so the processor skips

1618 - re-sampling (do_sample_frames=False) and keeps the original totals.

1619 - """

1614 + re-sampling (do_sample_frames=False) and keeps the original totals."""

1615 source = VideoSourceMetadata(total_frames_num=900, original_fps=30.0, duration=30.0)

1616 target = VideoTargetMetadata(num_frames=-1, fps=-1, max_duration=-1)

1617 indices = Glm5NextVideoBackend.compute_frames_index_to_sample(source, target)

@@ -1645,8 +1640,7 @@ def test_glm5next_backend_metadata_contract():

1640

1641 def _write_gray_video(tmp_path, total_frames, fps, size=(32, 32)):

1642 """Synthetic clip whose frame i is flat gray level i (near-lossless under

1648 - mp4v), so a decoded frame's level maps back to its source index.

1649 - """

1643 + mp4v), so a decoded frame's level maps back to its source index."""

1644 cv2 = pytest.importorskip("cv2")

1645

1646 path = tmp_path / f"gray_{total_frames}_{fps}.mp4"

@@ -1687,8 +1681,7 @@ class _CountingCap:

1681 def test_glm5next_backend_codec_parity(tmp_path, backend):

1682 """Every codec samples the same GLM indices and decodes the same

1683 frames; the OpenCV seek reader and torchcodec batched index-exact decode

1690 - must agree.

1691 - """

1684 + must agree."""

1685 if backend == "torchcodec":

1686 pytest.importorskip("torchcodec")

1687

@@ -1718,8 +1711,7 @@ def test_glm5next_backend_codec_parity(tmp_path, backend):

1711

1712 def test_glm5next_backend_decodes_only_sampled_frames(tmp_path):

1713 """End to end over a synthetic clip: load_bytes returns exactly the

1721 - sampler's frame count, with the right frame content at each index.

1722 - """

1714 + sampler's frame count, with the right frame content at each index."""

1715 pytest.importorskip("cv2")

1716

1717 total_frames, fps = 60, 10

@@ -1745,8 +1737,7 @@ def test_glm5next_backend_decodes_only_sampled_frames(tmp_path):

1737

1738 def test_glm5next_read_frames_seeks_past_large_gaps(tmp_path):

1739 """Sparse targets must not walk the container: the stock reader grabs

1748 - every frame up to the last index; the GLM reader seeks instead.

1749 - """

1740 + every frame up to the last index; the GLM reader seeks instead."""

1741 cv2 = pytest.importorskip("cv2")

1742

1743 total_frames, fps = 200, 10

@@ -1773,8 +1764,7 @@ def test_glm5next_read_frames_seeks_past_large_gaps(tmp_path):

1764

1765 def test_glm5next_read_frames_dense_walk_matches_stock(tmp_path):

1766 """Dense targets keep the sequential walk (seeking would be slower) and

1776 - return the same frames as the stock reader.

1777 - """

1767 + return the same frames as the stock reader."""

1768 cv2 = pytest.importorskip("cv2")

1769

1770 total_frames, fps = 120, 10

tests/parser/cohere/test_reasoning.py

+1 -2

@@ -1,8 +1,7 @@

1 # SPDX-License-Identifier: Apache-2.0

2 # SPDX-FileCopyrightText: Copyright contributors to the vLLM project

3 """Reasoning extraction, reasoning-end gating and parser selection for the

4 -unified Cohere Command parser.

5 -"""

4 +unified Cohere Command parser."""

5

6 from __future__ import annotations

7

tests/parser/cohere/utils.py

+3 -6

@@ -76,8 +76,7 @@ def token_deltas(

76 ) -> list[tuple[str, list[int]]]:

77 """Split ``text`` into ``(delta_text, delta_token_ids)`` steps of

78 ``chunk_size`` tokens, buffering incomplete multi-byte sequences (trailing

79 - U+FFFD) into the next step as real streaming does.

80 - """

79 + U+FFFD) into the next step as real streaming does."""

80 ids = tokenizer.encode(text)

81 deltas: list[tuple[str, list[int]]] = []

82 prev = ""

@@ -101,8 +100,7 @@ def make_parser(

100 model_config=None,

101 ) -> Parser:

102 """Resolve the unified Cohere parser through ``ParserManager`` by registry

104 - name (``cohere_command3`` / ``cohere_command4``).

105 - """

103 + name (``cohere_command3`` / ``cohere_command4``)."""

104 cls = ParserManager.get_parser(name, name, enable_auto_tools=True)

105 assert cls is not None

106 return cls(

@@ -119,8 +117,7 @@ def drive_parser(

117 deltas: list[tuple[str, list[int]]],

118 ) -> list[DeltaMessage]:

119 """Feed ``deltas`` through ``parse_delta`` (flushing on the last one) and

122 - return the non-``None`` messages in order.

123 - """

120 + return the non-``None`` messages in order."""

121 out: list[DeltaMessage] = []

122 for i, (text, ids) in enumerate(deltas):

123 delta = parser.parse_delta(text, ids, request, finished=i == len(deltas) - 1)

tests/parser/engine/test_deepseek_v32.py

+1 -2

@@ -289,8 +289,7 @@ class TestStreaming:

289

290 class TestMissingFunctionCallsWrapper:

291 """An invoke without the ``<｜DSML｜function_calls>`` wrapper is still a

292 - tool call (V3.2 counterpart of #48931).

293 - """

292 + tool call (V3.2 counterpart of #48931)."""

293

294 def test_non_streaming(self, mock_tokenizer, mock_request):

295 text = _invoke("fn", _param("k", "true", "v"))

tests/parser/engine/test_deepseek_v4.py

+6 -12

@@ -334,8 +334,7 @@ class TestMissingToolCallsWrapper:

334 ):

335 """Text after a tool block is dropped whether or not the closing

336 wrapper is present, so a missing ``</｜DSML｜tool_calls>`` does not

337 - change what the client sees.

338 - """

337 + change what the client sees."""

338 text = "pre\n" + self._ORPHAN

339 if trailing_end:

340 text += DSML_TOOL_END

@@ -916,8 +915,7 @@ class TestParallelUnwrapping:

915

916 class TestStreamingWrapperConsistency:

917 """Streamed arg deltas must stay consistent with final extraction

919 - when wrapper params like 'arguments' are unwrapped.

920 - """

918 + when wrapper params like 'arguments' are unwrapped."""

919

920 def test_streaming_wrapper_unwrap_consistency(self, mock_tokenizer, mock_request):

921 tool = _make_tool("get_weather", {"location": {"type": "string"}})

@@ -1198,8 +1196,7 @@ class TestDelegatingParserLargeDelta:

1196 )

1197 def test_eos_not_leaked_when_reasoning_never_ends(self, chunk_size):

1198 """EOS must not leak into reasoning_content when the model never

1201 - emits </think> (generation ends while still in REASONING state).

1202 - """

1199 + emits </think> (generation ends while still in REASONING state)."""

1200 eos_text = "<｜end▁of▁sentence｜>"

1201 eos_id = 128801

1202 vocab = {

@@ -1288,8 +1285,7 @@ class TestMalformedDsmlNoise:

1285 self, mock_tokenizer, mock_request, wrapper

1286 ):

1287 """Like the real wrapper, a corrupted one inside ``<think>`` closes

1291 - the reasoning block and starts the tool call.

1292 - """

1288 + the reasoning block and starts the tool call."""

1289 text = f"thinking{wrapper}\n{self._INNER}"

1290 parser = self._parser(mock_tokenizer, thinking=True)

1291 reasoning, content = parser.extract_reasoning(text, mock_request)

@@ -1333,8 +1329,7 @@ class TestMalformedDsmlNoise:

1329 prose mention of the real wrapper: the opener starts a tool block and

1330 the rest of the message is dropped. That is pre-existing behavior of

1331 ``<｜DSML｜tool_calls>``; the variants must not differ from it either

1336 - way.

1337 - """

1332 + way."""

1333 template = "The opener looks like {} and then params follow. Done."

1334 text = template.format(wrapper)

1335 reference = template.format(DSML_TOOL_START)

@@ -1358,8 +1353,7 @@ class TestMalformedDsmlNoise:

1353 @pytest.mark.parametrize("chunk_size", [1, 3, None], ids=lambda c: f"chunk={c}")

1354 def test_delegating_parser_corrupted_wrapper(self, wrapper, chunk_size):

1355 """The serving-layer shape: reasoning and tool adapters on separate

1361 - engines, with the corrupted opener arriving as plain text.

1362 - """

1356 + engines, with the corrupted opener arriving as plain text."""

1357 tokens = _dsv4_tokens(

1358 reasoning="Checking the weather.",

1359 tool_name="get_weather",

tests/parser/engine/test_delegating_replay.py

+1 -2

@@ -174,8 +174,7 @@ _TOOL_CALL_SAMPLES = [

174 )

175 def test_delegating_parse_tool_choice_none(parser_cls, parser_name, sample):

176 """Non-streaming parse() with tool_choice='none' via DelegatingParser

177 - must not leak special tokens into content.

178 - """

177 + must not leak special tokens into content."""

178 tokenizer = make_mock_tokenizer(sample)

179 validated_tools = (

180 _TOOLS_VALIDATOR.validate_python(sample.tools) if sample.tools else None

tests/parser/engine/test_engine.py

+19 -38

@@ -275,8 +275,7 @@ class TestStreaming:

275

276 def test_unmatched_close_brace_does_not_poison_depth(self):

277 """A stray } in malformed JSON must not kill streaming for

278 - all subsequent content.

279 - """

278 + all subsequent content."""

279 engine = StreamingParserEngine(_hermes_config(), tokenizer=None)

280 engine.feed("<tool_call>", [])

281

@@ -292,8 +291,7 @@ class TestStreaming:

291

292 def test_json_args_no_premature_close_brace(self):

293 """Closing braces of the top-level JSON shouldn't be streamed

295 - until confirmed by the end tag.

296 - """

294 + until confirmed by the end tag."""

295 engine = StreamingParserEngine(_hermes_config(), tokenizer=None)

296

297 engine.feed("<tool_call>", [])

@@ -348,8 +346,7 @@ class TestLexerBufferFlush:

346

347 def test_buffered_prefix_emitted_in_current_state(self):

348 """Text buffered by the lexer (e.g. '<') must be emitted as

351 - REASONING_CHUNK before THINK_END transitions to CONTENT.

352 - """

349 + REASONING_CHUNK before THINK_END transitions to CONTENT."""

350 engine = StreamingParserEngine(_think_config(), _make_think_tokenizer())

351

352 events = engine.feed("<think>", [_START_ID])

@@ -387,13 +384,11 @@ class TestLexerBufferFlush:

384

385 class TestTokenIdFiltering:

386 """When token IDs are available, lex-matched terminals that also

390 - have token_id_terminal entries should be demoted to content.

391 - """

387 + have token_id_terminal entries should be demoted to content."""

388

389 def test_lex_matched_terminal_demoted_after_token_ids_seen(self):

390 """After receiving token IDs, text that matches a token-ID

395 - terminal should be treated as content, not trigger a transition.

396 - """

391 + terminal should be treated as content, not trigger a transition."""

392 engine = StreamingParserEngine(_hermes_config(), _make_hermes_tokenizer())

393

394 # First feed with a non-special token ID to set _ever_had_token_ids

@@ -414,8 +409,7 @@ class TestTokenIdFiltering:

409

410 def test_scanner_matched_terminal_bypasses_filter(self):

411 """PreLexedTerminals from the scanner bypass the filter and

417 - still trigger state transitions.

418 - """

412 + still trigger state transitions."""

413 engine = StreamingParserEngine(_hermes_config(), _make_hermes_tokenizer())

414

415 events = engine.feed("<tool_call>", [_TOOL_START_ID])

@@ -428,8 +422,7 @@ class TestTokenIdFiltering:

422

423 def test_no_filtering_without_token_ids(self):

424 """When no token IDs are ever provided (non-streaming),

431 - text matching still triggers transitions.

432 - """

425 + text matching still triggers transitions."""

426 engine = StreamingParserEngine(_hermes_config(), _make_hermes_tokenizer())

427

428 events = engine.feed('<tool_call>{"name": "f"}</tool_call>', [])

@@ -441,8 +434,7 @@ class TestTokenIdFiltering:

434

435 def test_mixed_text_then_real_tool_call(self):

436 """Text mentioning tool syntax followed by a real special-token

444 - tool call.

445 - """

437 + tool call."""

438 engine = StreamingParserEngine(_hermes_config(), _make_hermes_tokenizer())

439

440 events1 = engine.feed("Mention <tool_call> in text. ", [1, 2, 3, 4])

@@ -502,8 +494,7 @@ class TestReasoningTokenCounts:

494

495 def _func_prefix_config() -> ParserEngineConfig:

496 """Config mixing token-ID terminals (TOOL_START/END) with

505 - text-only terminals (FUNC_PREFIX) and fallback transitions.

506 - """

497 + text-only terminals (FUNC_PREFIX) and fallback transitions."""

498 return ParserEngineConfig(

499 name="func_prefix_test",

500 terminals={

@@ -572,13 +563,11 @@ def _make_func_prefix_tokenizer():

563

564 class TestTextOnlyFallbackFiltering:

565 """When token IDs are available, transitions marked

575 - skip_in_token_id_mode should be skipped.

576 - """

566 + skip_in_token_id_mode should be skipped."""

567

568 def test_func_prefix_in_prose_demoted_in_strict_mode(self):

569 """<function=get_time> in prose should NOT trigger a tool call

580 - when strict mode is active.

581 - """

570 + when strict mode is active."""

571 engine = StreamingParserEngine(

572 _func_prefix_config(), _make_func_prefix_tokenizer()

573 )

@@ -595,8 +584,7 @@ class TestTextOnlyFallbackFiltering:

584

585 def test_normal_flow_after_tool_start_still_works(self):

586 """TOOL_START (special token) -> FUNC_PREFIX (text) should

598 - still parse a tool call normally in strict mode.

599 - """

587 + still parse a tool call normally in strict mode."""

588 engine = StreamingParserEngine(

589 _func_prefix_config(), _make_func_prefix_tokenizer()

590 )

@@ -616,8 +604,7 @@ class TestTextOnlyFallbackFiltering:

604

605 def test_fallback_fires_without_token_ids(self):

606 """When no token IDs are provided, fallback transitions should

619 - still fire normally.

620 - """

607 + still fire normally."""

608 engine = StreamingParserEngine(

609 _func_prefix_config(), _make_func_prefix_tokenizer()

610 )

@@ -631,8 +618,7 @@ class TestTextOnlyFallbackFiltering:

618

619 def test_tool_between_fallback_blocked_in_strict_mode(self):

620 """The (TOOL_BETWEEN, FUNC_PREFIX) fallback should also be

634 - blocked in strict mode.

635 - """

621 + blocked in strict mode."""

622 engine = StreamingParserEngine(

623 _func_prefix_config(), _make_func_prefix_tokenizer()

624 )

@@ -662,8 +648,7 @@ class TestNoUnusedTokenizerAttr:

648 class TestArgsResetOnReentry:

649 """When leaving TOOL_ARGS and later re-entering (e.g. two tool

650 calls), the entering-TOOL_ARGS block resets args tracking. The

665 - redundant reset on exit was removed.

666 - """

651 + redundant reset on exit was removed."""

652

653 @staticmethod

654 def _multi_tool_config() -> ParserEngineConfig:

@@ -738,8 +723,7 @@ class TestArgsResetOnReentry:

723

724 class TestToolPreambleFinish:

725 """finish() in TOOL_PREAMBLE state emits TOOL_CALL_END when a tool

741 - call was started (tool_index >= 0), but not when tool_index is -1.

742 - """

726 + call was started (tool_index >= 0), but not when tool_index is -1."""

727

728 @staticmethod

729 def _preamble_with_tool_call_start_config() -> ParserEngineConfig:

@@ -798,8 +782,7 @@ class TestToolPreambleFinish:

782

783 class TestRegexTerminalInfraRemoved:

784 """TerminalDef.priority, LexerShape.regex_terminals, and the regex

801 - matching loop were removed.

802 - """

785 + matching loop were removed."""

786

787 def test_terminal_def_no_priority(self):

788 import regex as re

@@ -824,8 +807,7 @@ class TestRegexTerminalInfraRemoved:

807

808 class TestMultiCharTerminalInArgs:

809 """Regression: multi-char terminals falling through in TOOL_ARGS

827 - must be fed char-by-char via _feed_args_text, not _feed_args_char.

828 - """

810 + must be fed char-by-char via _feed_args_text, not _feed_args_char."""

811

812 @staticmethod

813 def _newline_config() -> ParserEngineConfig:

@@ -917,8 +899,7 @@ class TestSkipToolParsing:

899

900 def _message_header_config(reasoning_end: bool) -> ParserEngineConfig:

901 """MESSAGE_HEADER-based grammar whose tool opener optionally carries

920 - REASONING_END, mirroring the two shapes a transition table can take.

921 - """

902 + REASONING_END, mirroring the two shapes a transition table can take."""

903 return ParserEngineConfig(

904 name=f"message_header_test_{reasoning_end}",

905 initial_state=ParserState.MESSAGE_HEADER,

tests/parser/engine/test_gemma4_streaming_reasoning.py

+9 -18

@@ -1097,8 +1097,7 @@ class TestNonStreamingReasoningPlusToolCalls:

1097

1098 def test_both_extractions_independent(self, parser, request_obj):

1099 """Calling extract_reasoning then extract_tool_calls on the same

1100 - parser instance should both work (each resets the engine).

1101 - """

1100 + parser instance should both work (each resets the engine)."""

1101 model_output = FULL_MODEL_OUTPUT

1102

1103 reasoning, _ = parser.extract_reasoning(model_output, request_obj)

@@ -1112,8 +1111,7 @@ class TestNonStreamingReasoningPlusToolCalls:

1111

1112 class TestAdapterExtractReasoning:

1113 """The reasoning adapter's extract_reasoning uses skip_tool_parsing

1115 - so tool call text is preserved as content for the tool adapter.

1116 - """

1114 + so tool call text is preserved as content for the tool adapter."""

1115

1116 @pytest.fixture

1117 def adapter(self, mock_tokenizer):

@@ -1154,8 +1152,7 @@ class TestAdapterExtractReasoning:

1152 class TestGemma4SchemaAwareTypeCoercion:

1153 """Verify that streaming and non-streaming produce identical

1154 type-fixed arguments when tool schemas declare string parameters

1157 - but the model outputs bare numbers/booleans.

1158 - """

1155 + but the model outputs bare numbers/booleans."""

1156

1157 @pytest.fixture

1158 def tools(self):

@@ -1185,8 +1182,7 @@ class TestGemma4SchemaAwareTypeCoercion:

1182

1183 def test_streaming_string_param_not_coerced(self, parser_with_tools, mock_request):

1184 """A numeric value for a string-typed param must remain a string

1188 - in the streamed output, matching the non-streaming result.

1189 - """

1185 + in the streamed output, matching the non-streaming result."""

1186 chunks = [

1187 "<|tool_call>",

1188 "call:update_record{",

@@ -1217,8 +1213,7 @@ class TestGemma4SchemaAwareTypeCoercion:

1213

1214 def test_streaming_matches_non_streaming(self, parser_with_tools, mock_request):

1215 """Concatenated streaming deltas must produce the same arguments

1220 - as non-streaming extraction.

1221 - """

1216 + as non-streaming extraction."""

1217 text = "<|tool_call>call:update_record{zipcode:12345}<tool_call|>"

1218

1219 non_streaming = parser_with_tools.extract_tool_calls(text, mock_request)

@@ -1239,8 +1234,7 @@ class TestGemma4SchemaAwareTypeCoercion:

1234

1235 class TestGemma4SchemaCoercionBoolNumberNull:

1236 """Verify that _fix_arg_types coerces string values to non-string

1242 - schema types for the Gemma4 parser.

1243 - """

1237 + schema types for the Gemma4 parser."""

1238

1239 @pytest.fixture

1240 def tools(self):

@@ -1301,8 +1295,7 @@ class TestGemma4SchemaCoercionBoolNumberNull:

1295

1296 def test_streaming_type_stability(self, parser_with_tools, mock_request):

1297 """Values streamed incrementally must not cause prefix

1304 - incompatibility when types are coerced.

1305 - """

1298 + incompatibility when types are coerced."""

1299 text = (

1300 "<|tool_call>call:configure{"

1301 "enabled:true,"

@@ -1402,8 +1395,7 @@ BARE_THOUGHT_SEQUENCE.append((TOOL_CALL_END_ID, "<tool_call|>"))

1395 class TestBareThoughtWithoutChannelOpener:

1396 """When the model omits <|channel> and starts with bare ``thought``,

1397 the parser should auto-inject the channel opener so reasoning is

1405 - captured correctly.

1406 - """

1398 + captured correctly."""

1399

1400 @pytest.fixture

1401 def bare_thought_tokenizer(self):

@@ -1474,8 +1466,7 @@ class TestBareThoughtWithoutChannelOpener:

1466

1467 def test_bare_thought_token_at_end_of_stream(self, request_obj):

1468 """When the stream ends with just "thought" (no \\n), the parser

1477 - should treat it as the thought prefix token, not real reasoning.

1478 - """

1469 + should treat it as the thought prefix token, not real reasoning."""

1470 seq: list[tuple[int, str]] = [

1471 (CHANNEL_START_ID, "<|channel>"),

1472 (3000, "thought"),

tests/parser/engine/test_glm47_moe.py

+2 -4

@@ -61,8 +61,7 @@ class TestIsReasoningEnd:

61

62 def test_previous_turn_tool_call_ignored(self, parser):

63 """A finished tool call from an earlier turn says nothing about the

64 - turn currently being generated.

65 - """

64 + turn currently being generated."""

65 prompt = [THINK_S, TEXT, TOOL_S, TEXT, TOOL_E, OBSERVATION, TEXT, ASSISTANT]

66 assert not parser.is_reasoning_end(prompt)

67

@@ -89,8 +88,7 @@ class TestExtractContentIds:

88

89 def test_falls_back_to_tool_call(self, parser):

90 """Without ``</think>``, content starts at the opener itself so the

92 - tool parser still receives a well-formed call.

93 - """

91 + tool parser still receives a well-formed call."""

92 ids = [THINK_S, TEXT, TOOL_S, 30, TOOL_E]

93 assert parser.extract_content_ids(ids) == [TOOL_S, 30, TOOL_E]

94

tests/parser/engine/test_inkling.py

+10 -20

@@ -70,8 +70,7 @@ _MARKERS = sorted(_TML_VOCAB, key=len, reverse=True)

70 def _tokenize(text: str) -> list[tuple[int, str]]:

71 """Tokenize like the real stream: markers are atomic special tokens,

72 plain text becomes one token per character (matching the mock

73 - tokenizer's ``chr``-based decode).

74 - """

73 + tokenizer's ``chr``-based decode)."""

74 tokens: list[tuple[int, str]] = []

75 i = 0

76 while i < len(text):

@@ -88,8 +87,7 @@ def _tokenize(text: str) -> list[tuple[int, str]]:

87

88 def _stream(parser, request, text: str, chunk_size: int):

89 """Stream production-shaped deltas: ``chunk_size`` tokens per delta,

91 - with delta_token_ids covering every token (specials and text).

92 - """

90 + with delta_token_ids covering every token (specials and text)."""

91 tokens = _tokenize(text)

92 results = []

93 previous_text = ""

@@ -121,8 +119,7 @@ def _stream(parser, request, text: str, chunk_size: int):

119 def _stream_text_only(parser, request, text: str, chunk_size: int):

120 """Stream text-only deltas (no token ids), chunked at arbitrary

121 character boundaries — exercises the text-lexing fallback path,

124 - including markers split across chunks.

125 - """

122 + including markers split across chunks."""

123 results = []

124 previous_text = ""

125 for start in range(0, len(text), chunk_size):

@@ -173,8 +170,7 @@ def _stream_delegating(parser, request, text, chunk_size, prompt_token_ids):

170 """Stream ``text`` through ``DelegatingParser.parse_delta``, ``chunk_size``

171 tokens per delta; return ``(content, reasoning, ordered tool names,

172 ordered tool arguments)``. Arguments arrive in fragments, so they are

176 - concatenated per tool index.

177 - """

173 + concatenated per tool index."""

174 tokens = _tokenize(text)

175 content, reasoning = "", ""

176 tools: dict[int, str] = {}

@@ -526,8 +522,7 @@ class TestToolCallFiltering:

522 """Inkling equivalents of the generic tool-call-filtering replay tests

523 (Inkling is excluded from those in test_replay.py: its structural

524 role/kind tokens and shared block-end token don't fit the generic

529 - reasoning/tool split model).

530 - """

525 + reasoning/tool split model)."""

526

527 def test_skip_tool_parsing_round_trip(self, mock_tokenizer, mock_request):

528 # First pass (reasoning adapter, skip_tool_parsing): reasoning is

@@ -825,8 +820,7 @@ class TestDelegatingTwoPass:

820 ):

821 """The optional function name between ``<|message_model|>`` and the

822 content-kind marker is metadata: the buffered header must be

828 - discarded on the way out, not flushed into content.

829 - """

823 + discarded on the way out, not flushed into content."""

824 tools = [_function_tool()]

825 mock_request.tools = tools

826 content, _, names, args = _stream_delegating(

@@ -843,8 +837,7 @@ class TestDelegatingTwoPass:

837 def test_content_state_tool_start_streaming(self, mock_tokenizer, mock_request):

838 """Same opener reached from CONTENT rather than MESSAGE_HEADER: a

839 text block closed with no ``<|message_model|>`` before the tool

846 - block. Preceding text must reach content exactly once, unmarked.

847 - """

840 + block. Preceding text must reach content exactly once, unmarked."""

841 tools = [_function_tool()]

842 mock_request.tools = tools

843 content, _, names, args = _stream_delegating(

@@ -866,8 +859,7 @@ def test_content_tool_start_emits_reasoning_end_in_reasoning_pass():

859 boundary (#49876), but the header-flush path

860 (MESSAGE_HEADER --END_MESSAGE--> CONTENT) reaches CONTENT without

861 one, so a tool block opening from there relies on this transition

869 - alone to hand off to the tool pass.

870 - """

862 + alone to hand off to the tool pass."""

863 engine = StreamingParserEngine(inkling_config(), tokenizer=None)

864 engine.skip_tool_parsing = True

865 engine.reset(initial_state=ParserState.CONTENT)

@@ -884,8 +876,7 @@ def test_skip_reasoning_parsing_inert_for_shared_markers():

876 markers. Inkling has none — ``<|end_message|>`` is labelled THINK_END

877 yet also closes text, header, and tool blocks — so the flag must be

878 inert: a tool block still closes through its transition and following

887 - text returns to CONTENT instead of leaking into the argument stream.

888 - """

879 + text returns to CONTENT instead of leaking into the argument stream."""

880 engine = StreamingParserEngine(inkling_config(), tokenizer=None)

881 engine.skip_reasoning_parsing = True

882 engine.reset(initial_state=ParserState.CONTENT)

@@ -905,8 +896,7 @@ class TestToolParserWithoutReasoningParser:

896 labels ``<|end_message|>`` as THINK_END; a bypass keyed on the label

897 would neutralize the closer of every block kind, leaking markers into

898 content and leaving tool calls unterminated. Structure must keep

908 - parsing exactly as with the reasoning parser attached.

909 - """

899 + parsing exactly as with the reasoning parser attached."""

900

901 GEN_PROMPT = [_TML_VOCAB[MSG_MODEL]]

902

tests/parser/engine/test_nemotron_v3.py

+6 -12

@@ -56,8 +56,7 @@ def parser():

56 class TestNemotronSwap:

57 def test_enable_thinking_false_swaps(self, parser):

58 """When enable_thinking=False, model output without think tags

59 - should have reasoning swapped to content.

60 - """

59 + should have reasoning swapped to content."""

60 text = "The answer is 42."

61 request = _make_request(enable_thinking=False)

62 reasoning, content = parser.extract_reasoning(text, request)

@@ -74,8 +73,7 @@ class TestNemotronSwap:

73

74 def test_no_swap_when_content_exists(self, parser):

75 """With enable_thinking=False but real </think> giving content,

77 - no swap occurs.

78 - """

76 + no swap occurs."""

77 text = "Some reasoning.</think>Actual content here."

78 request = _make_request(enable_thinking=False)

79 reasoning, content = parser.extract_reasoning(text, request)

@@ -182,13 +180,11 @@ class TestStreaming:

180

181 class TestParseDeltaTokenIdFiltering:

182 """parse_delta must not trigger tool call parsing when <tool_call>

185 - appears as regular text rather than as a special token ID.

186 - """

183 + appears as regular text rather than as a special token ID."""

184

185 def test_tool_call_text_in_reasoning_is_not_parsed(self, parser):

186 """Literal <tool_call> in model reasoning should be content,

190 - not a tool call.

191 - """

187 + not a tool call."""

188 request = _make_request()

189

190 text = (

@@ -214,8 +210,7 @@ class TestParseDeltaTokenIdFiltering:

210

211 def test_special_token_id_still_triggers_tool_call(self, parser):

212 """When the scanner matches a special token ID, the tool call

217 - must still be parsed correctly.

218 - """

213 + must still be parsed correctly."""

214 request = _make_request()

215

216 parser.parse_delta(

@@ -255,8 +250,7 @@ class TestParseDeltaTokenIdFiltering:

250

251 def test_text_discussion_then_real_tool_call(self, parser):

252 """Model discusses tool syntax in reasoning, then makes a real

258 - tool call via special tokens.

259 - """

253 + tool call via special tokens."""

254 request = _make_request()

255

256 r1 = parser.parse_delta(

tests/parser/engine/test_parser_engine.py

+23 -46

@@ -323,8 +323,7 @@ class TestEventsToDelta:

323

324 def test_multiple_arg_chunks_same_batch_coalesced(self):

325 """Multiple events for the same tool in one batch must produce

326 - at most one DeltaToolCall per index.

327 - """

326 + at most one DeltaToolCall per index."""

327 engine = _make_engine()

328 events = [

329 SemanticEvent(EventType.TOOL_CALL_START, tool_index=0),

@@ -552,8 +551,7 @@ class TestPostToolContentDeferral:

551 """Regression: content after TOOL_CALL_END in the same batch must not

552 produce a mixed DeltaMessage(content=..., tool_calls=...) — that causes

553 split_delta to reorder content before tool_calls, breaking the Responses

555 - API state machine.

556 - """

554 + API state machine."""

555

556 def test_text_after_tool_end_deferred(self):

557 engine = _make_engine()

@@ -607,8 +605,7 @@ class TestPostToolContentDeferral:

605 """Deferred content from batch N must not mix with arg-continuation

606 tool events in batch N+1 — that creates a DeltaMessage with both

607 content and nameless tool_calls, which crashes the Responses API

610 - state machine (name=None → Pydantic ValidationError).

611 - """

608 + state machine (name=None → Pydantic ValidationError)."""

609 engine = _make_engine()

610 engine._content_has_nonws = True

611

@@ -837,8 +834,7 @@ class TestBuildExtractedResult:

834

835 class TestEngineBasedPath:

836 """Tests for the _engine_based accumulation behavior in

840 - DelegatingParser.parse_delta.

841 - """

837 + DelegatingParser.parse_delta."""

838

839 def test_engine_based_true_when_both_parsers_engine(self):

840 r = SimpleNamespace(engine_based_streaming=True)

@@ -884,8 +880,7 @@ class TestEngineBasedPath:

880

881 class TestParseTokenIdPassthrough:

882 """parse() must forward model_output_token_ids to _single_pass_parse

887 - so that token-ID-based strict terminal matching is active.

888 - """

883 + so that token-ID-based strict terminal matching is active."""

884

885 def test_literal_tool_tag_in_content_preserved_with_token_ids(self, mock_request):

886 engine = _make_engine(_hermes_config())

@@ -1109,8 +1104,7 @@ class TestAdapterFinishOnStreamEnd:

1104

1105 def test_lexer_buffer_flushed_on_finished(self):

1106 """Text buffered as a potential terminal prefix must be emitted

1112 - as content when the stream ends.

1113 - """

1107 + as content when the stream ends."""

1108 tokenizer = make_mock_tokenizer(_VOCAB)

1109 parser = _CombinedDelegating(tokenizer)

1110 request = _make_delegating_request()

@@ -1128,8 +1122,7 @@ class TestAdapterFinishOnStreamEnd:

1122

1123 def test_args_buffer_flushed_on_finished(self):

1124 """Pending arg buffer text must be emitted when stream ends

1131 - mid-tool-call (closing brace held back in buffer).

1132 - """

1125 + mid-tool-call (closing brace held back in buffer)."""

1126 tokenizer = make_mock_tokenizer(_VOCAB)

1127 parser = _CombinedDelegating(tokenizer)

1128 request = _make_delegating_request()

@@ -1365,8 +1358,7 @@ class TestSkipToolSpanForwarding:

1358

1359 class TestToolAdapterForwardsKwargs:

1360 """ParserEngineToolAdapter.__init__ must forward **kwargs to the

1368 - parser engine class so chat_template_kwargs reach model parsers.

1369 - """

1361 + parser engine class so chat_template_kwargs reach model parsers."""

1362

1363 @pytest.mark.parametrize(

1364 "enable_thinking,expected_state",

@@ -1396,8 +1388,7 @@ class TestToolAdapterForwardsKwargs:

1388

1389 class TestExtractContentIdsNoEmptyReturn:

1390 """extract_content_ids must return input_ids (not []) when there is

1399 - no THINK_END token ID and _reasoning_ended is True.

1400 - """

1391 + no THINK_END token ID and _reasoning_ended is True."""

1392

1393 _NO_THINK_CONFIG = ParserEngineConfig(name="no_think_end", token_id_terminals={})

1394

@@ -1795,8 +1786,7 @@ class TestDropSpecialTokens:

1786

1787 def test_drops_special_token_by_id_from_content(self):

1788 """A special token (not a configured terminal) is dropped when

1798 - it arrives as its actual token ID.

1799 - """

1789 + it arrives as its actual token ID."""

1790 config = ParserEngineConfig(

1791 name="drop_content_test",

1792 terminals={},

@@ -1831,8 +1821,7 @@ class TestDropSpecialTokens:

1821

1822 def test_drops_via_text_fallback_when_no_token_ids(self):

1823 """When no token IDs are provided, text-based lexer catches

1834 - drop tokens as a fallback.

1835 - """

1824 + drop tokens as a fallback."""

1825 engine = _make_engine(

1826 vocab=_DROP_VOCAB,

1827 special_tokens=list(_DROP_VOCAB.keys()),

@@ -1845,8 +1834,7 @@ class TestDropSpecialTokens:

1834

1835 def test_regular_tokens_spelling_special_survive(self):

1836 """Regular tokens that spell out a drop-token string survive

1848 - when token IDs prove they are not the special token.

1849 - """

1837 + when token IDs prove they are not the special token."""

1838 vocab = {**_DROP_VOCAB, "h": 72, "<": 73, "bos": 74, ">": 75, "w": 76}

1839 engine = _make_engine(

1840 vocab=vocab,

@@ -1862,8 +1850,7 @@ class TestDropSpecialTokens:

1850

1851 def test_configured_terminal_not_treated_as_drop(self):

1852 """Tokens already in config.terminals (like <think>) are handled

1865 - by the state machine, not the drop mechanism.

1866 - """

1853 + by the state machine, not the drop mechanism."""

1854 engine = _make_engine(

1855 vocab=_DROP_VOCAB,

1856 special_tokens=list(_DROP_VOCAB.keys()),

@@ -1918,8 +1905,7 @@ class TestDropSpecialTokens:

1905 def test_drops_applied_with_skip_tool_parsing(self):

1906 """Drop tokens are always dropped, even with skip_tool_parsing.

1907 DROP_TERMINALs have no transitions by construction, so no parser

1921 - pass can use them.

1922 - """

1908 + pass can use them."""

1909 for initial_state in (ParserState.REASONING, ParserState.CONTENT):

1910 engine = _make_engine(

1911 vocab=_DROP_VOCAB,

@@ -1935,8 +1921,7 @@ class TestDropSpecialTokens:

1921

1922 def test_transitions_unaffected_by_drop_in_reasoning_with_skip_tool_parsing(self):

1923 """With skip_tool_parsing in REASONING state, drop tokens are

1938 - removed but configured terminals still fire their transitions.

1939 - """

1924 + removed but configured terminals still fire their transitions."""

1925 engine = _make_engine(

1926 vocab=_DROP_VOCAB,

1927 special_tokens=list(_DROP_VOCAB.keys()),

@@ -1972,8 +1957,7 @@ class TestDropSpecialTokens:

1957

1958 def test_mixed_configured_and_drop_terminals(self):

1959 """Configured terminals trigger transitions while drop terminals

1975 - are silently removed in the same stream.

1976 - """

1960 + are silently removed in the same stream."""

1961 engine = _make_engine(

1962 vocab=_DROP_VOCAB,

1963 special_tokens=list(_DROP_VOCAB.keys()),

@@ -1997,8 +1981,7 @@ class TestTruncatedToolOpenerStreamParity:

1981 """Regression tests for #47137: when generation terminates inside a

1982 ``<tool_call>`` opener that has not been promoted to a tool call

1983 (via ``max_tokens`` or a ``stop`` string), the non-streaming path

2000 - must drop the incomplete markup, matching the streaming path.

2001 - """

1984 + must drop the incomplete markup, matching the streaming path."""

1985

1986 _QWEN3_VOCAB = {

1987 "<tool_call>": 100,

@@ -2044,8 +2027,7 @@ class TestTruncatedToolOpenerStreamParity:

2027 )

2028 def test_truncated_opener_dropped_in_both_paths(self, mock_request, chunks):

2029 """A cutoff inside the opener yields no content and no tool calls

2047 - in both the non-streaming and streaming paths.

2048 - """

2030 + in both the non-streaming and streaming paths."""

2031 text = "".join(chunks)

2032

2033 _, content, tool_calls = self._make_parser().parse(

@@ -2058,8 +2040,7 @@ class TestTruncatedToolOpenerStreamParity:

2040

2041 def test_content_before_truncated_opener_preserved(self, mock_request):

2042 """Only the incomplete markup is dropped; content generated before

2061 - the opener is returned identically by both paths.

2062 - """

2043 + the opener is returned identically by both paths."""

2044 chunks = ["Checking the weather. ", "<tool_call>", "\n", "<function"]

2045 text = "".join(chunks)

2046

@@ -2073,8 +2054,7 @@ class TestTruncatedToolOpenerStreamParity:

2054

2055 def test_content_around_unpromoted_tool_block_stays_ordered(self, mock_request):

2056 """Text surrounding a complete-but-unparsable tool block keeps its

2076 - original order in the non-streaming path, matching streaming.

2077 - """

2057 + original order in the non-streaming path, matching streaming."""

2058 chunks = ["A ", "<tool_call>", "garbage", "</tool_call>", " B"]

2059 text = "".join(chunks)

2060

@@ -2088,8 +2068,7 @@ class TestTruncatedToolOpenerStreamParity:

2068

2069 def test_complete_tool_call_still_promoted(self, mock_request):

2070 """Sanity check: a complete tool call still parses in the

2091 - non-streaming path after the truncation fix.

2092 - """

2071 + non-streaming path after the truncation fix."""

2072 text = (

2073 "<tool_call>\n"

2074 "<function=get_weather>\n"

@@ -2109,8 +2088,7 @@ class TestThinkMarkupWithoutReasoningParser:

2088 """With no reasoning parser configured, reasoning markup is plain

2089 content: engine-backed tool parsers must return it verbatim in both

2090 the non-streaming and streaming paths, not consume the markers or

2112 - reclassify the block as reasoning and drop it.

2113 - """

2091 + reclassify the block as reasoning and drop it."""

2092

2093 _VOCAB = {

2094 "<think>": 90,

@@ -2176,8 +2154,7 @@ class TestThinkMarkupWithoutReasoningParser:

2154

2155 def test_think_block_preserved_alongside_tool_call(self, mock_request):

2156 """The tools-called branch also returns the tool parser's content;

2179 - a think block preceding a promoted tool call must survive in it.

2180 - """

2157 + a think block preceding a promoted tool call must survive in it."""

2158 think = "<think>plan</think>"

2159 text = think + (

2160 "<tool_call>\n"

tests/parser/engine/test_qwen3.py

+5 -10

@@ -318,8 +318,7 @@ class TestStreaming:

318

319 def test_streaming_args_arrive_incrementally(self, parser, mock_request):

320 """Arguments must stream as intermediate deltas, not batch at

321 - tool-end.

322 - """

321 + tool-end."""

322 chunks = [

323 "<tool_call>\n",

324 "<function=get_weather>\n",

@@ -524,8 +523,7 @@ class TestStreaming:

523

524 def test_streaming_split_tool_call_tag(self, parser, mock_request):

525 """<tool_call> arrives as a single special token; the rest of

527 - the content is split into fine-grained chunks.

528 - """

526 + the content is split into fine-grained chunks."""

527 chunks = [

528 "<tool_call>\n",

529 "<function=test>\n",

@@ -702,8 +700,7 @@ class TestArgConverter:

700

701 class TestSchemaAwareTypeCoercion:

702 """Verify that _fix_arg_types corrects miscoerced values using the

705 - tool schema.

706 - """

703 + tool schema."""

704

705 @pytest.fixture

706 def tools(self):

@@ -845,8 +842,7 @@ class TestAnyOfTypeCoercion:

842

843 def test_anyof_string_param_not_coerced(self, parser_with_anyof, mock_request):

844 """A param with anyOf including 'string' must not be coerced

848 - to integer.

849 - """

845 + to integer."""

846 text = (

847 "<tool_call>\n"

848 "<function=set_config>\n"

@@ -862,8 +858,7 @@ class TestAnyOfTypeCoercion:

858

859 class TestSchemaCoercionBoolNumberNull:

860 """Verify that _fix_arg_types coerces string values to non-string

865 - schema types using coerce_to_schema_type.

866 - """

861 + schema types using coerce_to_schema_type."""

862

863 @pytest.fixture

864 def tools(self):

tests/parser/engine/test_qwen3_reasoning.py

+6 -12

@@ -276,8 +276,7 @@ class TestIsReasoningEndTurnBoundaries:

276

277 def test_empty_boundary_config_keeps_global_walk(self):

278 """Configs without turn_boundary_tokens keep the pre-existing

279 - behavior: any </think> in the sequence ends reasoning.

280 - """

279 + behavior: any </think> in the sequence ends reasoning."""

280 cfg = qwen3_config(turn_boundary_tokens=frozenset())

281 parser = Qwen3Parser(

282 make_mock_tokenizer(_QWEN3_VOCAB), parser_engine_config=cfg

@@ -289,8 +288,7 @@ class TestIsReasoningEndTurnBoundaries:

288 def test_boundary_tokens_missing_from_vocab_keep_global_walk(self):

289 """Boundary tokens absent from the vocabulary resolve to nothing,

290 so subclasses reusing the default on a non-ChatML vocab keep the

292 - pre-existing behavior.

293 - """

291 + pre-existing behavior."""

292 vocab = {

293 token: token_id

294 for token, token_id in _QWEN3_VOCAB.items()

@@ -301,8 +299,7 @@ class TestIsReasoningEndTurnBoundaries:

299

300 def test_boundary_with_thinking_disabled_is_end(self, mock_tokenizer):

301 """With thinking disabled (initial state CONTENT) a walk that stops at a turn

304 - boundary must report reasoning as ended: the model never emits a marker.

305 - """

302 + boundary must report reasoning as ended: the model never emits a marker."""

303 parser = Qwen3Parser(

304 mock_tokenizer, chat_template_kwargs={"enable_thinking": False}

305 )

@@ -310,8 +307,7 @@ class TestIsReasoningEndTurnBoundaries:

307

308 def test_boundary_with_thinking_enabled_not_end(self, parser):

309 """With thinking disabled (initial state REASONING) a walk that stops at a turn

313 - boundary must keep reasoning open

314 - """

310 + boundary must keep reasoning open"""

311 assert not parser.is_reasoning_end([_IM_START_ID, _TEXT_ID])

312

313

@@ -672,8 +668,7 @@ class TestTrailingWhitespaceStripping:

668

669 class TestWhitespaceStrippingDisabled:

670 """When strip_trailing_reasoning_whitespace is False,

675 - trailing whitespace in reasoning must be preserved.

676 - """

671 + trailing whitespace in reasoning must be preserved."""

672

673 @pytest.fixture

674 def parser_no_strip(self):

@@ -731,8 +726,7 @@ class TestThinkingDisabled:

726 def test_thinking_disabled_streaming_content_only(self, mock_tokenizer):

727 """Plain text with thinking disabled must stream as content, not

728 reasoning. Before the fix, the REASONING initial state caused all

734 - output to be emitted as reasoning chunks.

735 - """

729 + output to be emitted as reasoning chunks."""

730 p = Qwen3Parser(

731 mock_tokenizer,

732 chat_template_kwargs={"enable_thinking": False},

tests/parser/engine/test_replay.py

+4 -8

@@ -440,8 +440,7 @@ class TestToolCallFilteringReplay:

440 )

441 class TestToolCallFilteringNonStreaming:

442 """Non-streaming parse() with tool_choice='none' must suppress tool

443 - calls and not leak special tokens into content.

444 - """

443 + calls and not leak special tokens into content."""

444

445 def test_parse(self, parser_cls, sample, think_end, tool_start):

446 tokenizer = make_mock_tokenizer(sample)

@@ -484,8 +483,7 @@ _WS_TOOL_SAMPLES = [(t[0], t[1]) for t in _TOOL_CALL_SAMPLES if "whitespace" in

483 )

484 class TestToolChoiceNoneStreamingParity:

485 """Streaming and non-streaming must return the same content

487 - when tool_choice='none' suppresses tool calls.

488 - """

486 + when tool_choice='none' suppresses tool calls."""

487

488 def test_content_matches(self, parser_cls, sample):

489 tokenizer = make_mock_tokenizer(sample)

@@ -543,8 +541,7 @@ def _inject_drop_tokens(sample):

541

542 class TestDropTokenReplay:

543 """Verify unconfigured special tokens are silently dropped across

546 - all parsers and chunk sizes.

547 - """

544 + all parsers and chunk sizes."""

545

546 @pytest.mark.parametrize(

547 "parser_info",

@@ -601,8 +598,7 @@ class TestDropTokenNonStreaming:

598

599 class TestAdapterReferences:

600 """Verify make_adapters sets reasoning/tool parser class refs on parser engine

604 - parser classes so the serving layer finds them and calls adjust_request.

605 - """

601 + parser classes so the serving layer finds them and calls adjust_request."""

602

603 @pytest.mark.parametrize(

604 "parser_name",

tests/parser/engine/test_seed_oss.py

+2 -4

@@ -129,8 +129,7 @@ def test_single_tool_call(tool_parser, mock_request):

129

130 def test_malformed_function_end_does_not_drop_siblings(tool_parser, mock_request):

131 """Regression for #46314: a malformed ``</function>`` with no closing ``>``

132 - on the header must not discard the other, well-formed calls.

133 - """

132 + on the header must not discard the other, well-formed calls."""

133 text = (

134 f"{TOOL_CALL_START}\n<function=broken</function>\n{TOOL_CALL_END}"

135 f"{TOOL_CALL_START}\n<function=get_weather>\n"

@@ -172,8 +171,7 @@ def test_reasoning_then_tool_call(parser):

171

172 def test_streaming_think_end_and_tool_call_same_delta(parser):

173 """``</seed:think>`` and ``<seed:tool_call>`` arriving in one delta must

175 - not leak the terminal tokens into the reasoning text.

176 - """

174 + not leak the terminal tokens into the reasoning text."""

175 reasoning, content = simulate_reasoning_streaming(

176 parser,

177 [

tests/parser/engine/test_ufffd_reasoning_transition.py

+1 -2

@@ -146,8 +146,7 @@ class TestUfffdReasoningTransition:

146

147 def test_byte_fallback_tokenizer_produces_ufffd(self):

148 """Validate the fixture: decode() returns U+FFFD for isolated

149 - byte-fallback token IDs, proving the old code path would leak.

150 - """

149 + byte-fallback token IDs, proving the old code path would leak."""

150 vocab = dict(_MODEL_CONFIGS[0].values[0])

151 tokenizer = ByteFallbackMockTokenizer(vocab, _SHARED_TOKENS, _SHARED_UFFFD_IDS)

152 assert tokenizer.decode([200]) == "�"

tests/parser/engine/trace_builder.py

+1 -2

@@ -992,8 +992,7 @@ def _inkling_block(

992 ) -> None:

993 """Append one Inkling content block; blocks after the first start with

994 the ``<|message_model|>`` role token (the first block continues the

995 - generation prompt directly).

996 - """

995 + generation prompt directly)."""

996 if segs:

997 segs.append(("<|message_model|>", True))

998 segs.append((kind_token, True))

tests/parser/mistral/test_tool_calls.py

+5 -10

@@ -492,8 +492,7 @@ def test_extract_tool_calls_pre_v11_regex_fallback(

492 ):

493 """The regex fallback path finds valid JSON via regex when the primary

494 raw_decode fails on leading junk. It should re-serialize arguments

495 - and return a valid tool call.

496 - """

495 + and return a valid tool call."""

496 model_output = (

497 '[TOOL_CALLS] junk [{"name": "add", "arguments":{"a": 1, "b": 2}}] trail'

498 )

@@ -2004,8 +2003,7 @@ def test_streaming_pre_v11_parallel_calls_batched_deltas(

2003 mistral_pre_v11_tool_parser, mistral_pre_v11_tokenizer, chunk_size, driver

2004 ):

2005 """A batched delta spanning the boundary between two parallel calls must

2007 - keep them on distinct indices (the bug collapsed both onto index 0).

2008 - """

2006 + keep them on distinct indices (the bug collapsed both onto index 0)."""

2007 model_output = (

2008 '[TOOL_CALLS] [{"name": "add", "arguments": {"a": 3.5, "b": 4}}, '

2009 '{"name": "get_current_weather", "arguments": '

@@ -2123,8 +2121,7 @@ def test_reasoning_active_no_think_block_no_leak(

2121 chunk_size,

2122 ):

2123 """Tool call without a preceding think block must not leak as content

2126 - when the reasoning parser is active.

2127 - """

2124 + when the reasoning parser is active."""

2125 accumulated_content = ""

2126 function_names: list[str] = []

2127 function_args_strs: list[str] = []

@@ -2225,8 +2222,7 @@ def test_adjust_request_pre_v11_required_clears_response_format(

2222 response_format so the tool schema is the sole structured-output

2223 constraint, matching the base ToolParser (otherwise the request either

2224 hits the "multiple constraints" engine error or, with no constraint,

2228 - rambles to finish_reason='length').

2229 - """

2225 + rambles to finish_reason='length')."""

2226 request = _make_request(tool_choice="required", response_format=response_format)

2227 result = mistral_pre_v11_tool_parser.adjust_request(request)

2228

@@ -2243,8 +2239,7 @@ def test_adjust_request_non_mistral_tokenizer_required_injects_schema(

2239 ) -> None:

2240 """The guided-schema injection must also fire for non-Mistral (e.g. HF-mode)

2241 tokenizers driving the Mistral tool parser, mirroring the base ToolParser so

2246 - required + response_format does not fall back to an unconstrained ramble.

2247 - """

2242 + required + response_format does not fall back to an unconstrained ramble."""

2243 request = _make_request(

2244 tool_choice="required", response_format={"type": "json_object"}

2245 )

tests/parser/test_streaming.py

+12 -24

@@ -229,8 +229,7 @@ def _boundary_chunks(tokenizer, parser, end_token_id=None):

229

230 def test_parse_delta_reasoning_not_dropped_on_boundary(tokenizer, request_obj):

231 """Regression: reasoning must not be lost when a multi-token delta

232 - spans the reasoning/tool-call boundary.

233 - """

232 + spans the reasoning/tool-call boundary."""

233 parser = make_parser(tokenizer, reasoning=True, tool=True)

234 chunks = _boundary_chunks(tokenizer, parser)

235 results = stream_chunks(parser, tokenizer, chunks, request_obj)

@@ -248,8 +247,7 @@ def test_parse_delta_reasoning_not_dropped_on_boundary(tokenizer, request_obj):

247

248 def test_parse_delta_reasoning_boundary_no_tool_parser(tokenizer, request_obj):

249 """When no tool parser is active, boundary-spanning chunks must still

251 - preserve reasoning and pass post-</think> text as content.

252 - """

250 + preserve reasoning and pass post-</think> text as content."""

251 parser = make_parser(tokenizer, reasoning=True, tool=False)

252 chunks = _boundary_chunks(tokenizer, parser)

253 results = stream_chunks(parser, tokenizer, chunks, request_obj)

@@ -263,8 +261,7 @@ def test_parse_delta_reasoning_boundary_no_tool_parser(tokenizer, request_obj):

261

262 def test_parse_delta_reasoning_only_no_think_leak(tokenizer, request_obj):

263 """Regression: </think> must not leak into content when streaming

266 - token-by-token with reasoning=True, tool=False.

267 - """

264 + token-by-token with reasoning=True, tool=False."""

265 parser = make_parser(tokenizer, reasoning=True, tool=False)

266 results = stream_text(

267 parser, tokenizer, MODEL_OUTPUT, request_obj, prompt_token_ids=[]

@@ -306,8 +303,7 @@ def test_parse_delta_reasoning_only_thinking_disabled(tokenizer, request_obj):

303

304 def test_parse_delta_finished_no_flush_without_tool_call_delta(tokenizer, request_obj):

305 """When finished=True but the final parse_delta produces no

309 - tool-call delta, unstreamed args are not flushed.

310 - """

306 + tool-call delta, unstreamed args are not flushed."""

307 parser = make_parser(tokenizer, reasoning=False, tool=True)

308

309 results = stream_text(

@@ -330,8 +326,7 @@ def test_parse_delta_finished_no_flush_without_tool_call_delta(tokenizer, reques

326

327 def test_parse_delta_finished_no_extra_args_when_fully_streamed(tokenizer, request_obj):

328 """When all args have been streamed, finished=True must not

333 - produce extra or duplicate arguments.

334 - """

329 + produce extra or duplicate arguments."""

330 parser = make_parser(tokenizer, reasoning=False, tool=True)

331 results = stream_text(

332 parser, tokenizer, MODEL_OUTPUT, request_obj, prompt_token_ids=[]

@@ -351,8 +346,7 @@ def test_parse_delta_finished_no_extra_args_when_fully_streamed(tokenizer, reque

346

347 def test_parse_delta_finished_appends_remaining_args(tokenizer, request_obj):

348 """When finished=True and the tool parser has unstreamed args,

354 - parse_delta appends the remaining arguments to the tool-call delta.

355 - """

349 + parse_delta appends the remaining arguments to the tool-call delta."""

350 parser = make_parser(tokenizer, reasoning=False, tool=True)

351 token_ids = tokenizer.encode(MODEL_OUTPUT, add_special_tokens=False)

352

@@ -499,8 +493,7 @@ def test_engine_reasoning_hermes_tool_token_by_token(tokenizer, request_obj):

493 """Qwen3 engine reasoning + Hermes tool parser, token-by-token.

494

495 Sanity check that the mixed engine/non-engine configuration works

502 - when tokens arrive one at a time (no deferred content).

503 - """

496 + when tokens arrive one at a time (no deferred content)."""

497 parser = Qwen3ReasoningHermesToolParser(tokenizer)

498

499 assert parser._reasoning_parser.engine_based_streaming is True

@@ -527,8 +520,7 @@ def test_engine_reasoning_hermes_tool_boundary(tokenizer, request_obj):

520

521 When </think> and <tool_call> are in the same chunk with aligned

522 text and token IDs, the engine processes both terminals and returns

530 - the <tool_call> text as content.

531 - """

523 + the <tool_call> text as content."""

524 parser = Qwen3ReasoningHermesToolParser(tokenizer)

525 end_token_id = parser._reasoning_parser._parser_engine._reasoning_end_token_id

526 chunks = _boundary_chunks(tokenizer, parser, end_token_id=end_token_id)

@@ -559,8 +551,7 @@ def test_engine_reasoning_hermes_tool_text_holdback(tokenizer, request_obj):

551 Without the fix, finish_streaming() is never called at the

552 reasoning->tool transition when _engine_based is False, so the '<'

553 is lost and the Hermes parser sees 'tool_call>...' instead of

562 - '<tool_call>...'.

563 - """

554 + '<tool_call>...'."""

555 parser = Qwen3ReasoningHermesToolParser(tokenizer)

556 vocab = tokenizer.get_vocab()

557 think_end_id = vocab["</think>"]

@@ -651,8 +642,7 @@ def test_engine_reasoning_no_tool_batched_content_passthrough(tokenizer, request

642 batched delta carries ``</think>`` plus the following content

643 (as happens with stream_interval > 1). The post-``</think>`` text

644 must be emitted as content -- not dropped, not reclassified as

654 - reasoning -- and the ``</think>`` marker must not leak either way.

655 - """

645 + reasoning -- and the ``</think>`` marker must not leak either way."""

646 parser = Qwen3ReasoningNoToolParser(tokenizer)

647 assert parser._reasoning_parser.engine_based_streaming is True

648 assert parser._tool_parser is None

@@ -680,8 +670,7 @@ def _decode_stream_deltas(tokenizer, groups):

670 This mirrors how vLLM's detokenizer feeds ``parse_delta`` in

671 production: byte-level UTF-8 hold-back means a character whose bytes

672 span multiple tokens is only surfaced once complete (a naive

683 - per-token ``decode`` would instead emit U+FFFD replacement chars).

684 - """

673 + per-token ``decode`` would instead emit U+FFFD replacement chars)."""

674 from tokenizers.decoders import DecodeStream

675

676 stream = DecodeStream(skip_special_tokens=False)

@@ -713,8 +702,7 @@ def test_engine_reasoning_hermes_tool_multibyte_holdback(tokenizer, request_obj)

702 ``DecodeStream``. Its UTF-8 hold-back yields the correct

703 character round-trip (a naive per-token decode would corrupt it),

704 verifying the boundary stays byte-safe for multi-token

716 - characters.

717 - """

705 + characters."""

706 parser = Qwen3ReasoningHermesToolParser(tokenizer)

707 vocab = tokenizer.get_vocab()

708 think_end_id = vocab["</think>"]

tests/plugins_tests/test_endpoint_plugins.py

+4 -8

@@ -81,8 +81,7 @@ def test_no_plugins_loaded_when_allowlist_is_empty_string(

81 ):

82 """`VLLM_PLUGINS=""` parses to `[""]`, not `None` (see `vllm.envs`), so it

83 must be treated as a (non strict) allowlist matching no plugin name, not

84 - as "unset".

85 - """

84 + as "unset"."""

85 monkeypatch.setenv("VLLM_PLUGINS", "")

86

87 assert load_endpoint_plugins(("generate",)) == []

@@ -167,8 +166,7 @@ def test_attach_is_noop_when_nothing_discovered(monkeypatch: pytest.MonkeyPatch)

166 async def test_init_state_is_noop_without_phase_a(monkeypatch: pytest.MonkeyPatch):

167 """`init_app_state` callers that never ran `build_app` (e.g.

168 `run_batch.py`, which builds a bare `State()`) must not crash just

170 - because `state.endpoint_plugins` was never set.

171 - """

169 + because `state.endpoint_plugins` was never set."""

170 from starlette.datastructures import State

171

172 monkeypatch.setenv("VLLM_PLUGINS", "dummy_admin_endpoint_plugin")

@@ -186,8 +184,7 @@ async def test_render_server_attaches_endpoint_plugins_with_no_engine_client(

184 """The CPU only render server has no `EngineClient` but a plugin eligible

185 for the `render` task (`required_tasks` is `None` or includes `"render"`)

186 still gets its routes attached at Phase A. Phase B passes `None` for

189 - `engine_client` and it's up to the plugin to handle that.

190 - """

187 + `engine_client` and it's up to the plugin to handle that."""

188 monkeypatch.setenv("VLLM_PLUGINS", "dummy_admin_endpoint_plugin")

189

190 args = _build_args()

@@ -213,8 +210,7 @@ async def test_render_server_attaches_endpoint_plugins_with_no_engine_client(

210 @pytest.mark.asyncio

211 async def test_endpoint_plugin_end_to_end(monkeypatch: pytest.MonkeyPatch):

212 """Phase A (attach) + Phase B (init) wired through `build_app` then

216 - exercised with a real HTTP request against the worked example plugin.

217 - """

213 + exercised with a real HTTP request against the worked example plugin."""

214 monkeypatch.setenv("VLLM_PLUGINS", "dummy_admin_endpoint_plugin")

215

216 args = _build_args()

tests/plugins_tests/test_io_processor_plugins.py

+1 -2

@@ -38,8 +38,7 @@ def my_plugin_entry_points():

38 """Patch importlib.metadata.entry_points to expose a single 'my_plugin'

39 entry point backed by DummyIOProcessor, exercising the full plugin-loading

40 code path: entry_points → plugin.load() → func() →

41 - resolve_obj_by_qualname → IOProcessor.__init__.

42 - """

41 + resolve_obj_by_qualname → IOProcessor.__init__."""

42 qualname = f"{DummyIOProcessor.__module__}.{DummyIOProcessor.__qualname__}"

43 ep = MagicMock()

44 ep.name = "my_plugin"

tests/quantization/test_auto_round.py

+5 -10

@@ -382,8 +382,7 @@ def test_inc_get_quant_method_unquantized_layer_with_model_prefix(

382 layer_factory,

383 ) -> None:

384 """``model.``-prefixed extra_config entries trigger early exit for both

385 - LinearBase and ParallelLMHead layers.

386 - """

385 + LinearBase and ParallelLMHead layers."""

386 config = make_config(extra_config={"model.layer": {"bits": 16}})

387 layer = layer_factory()

388

@@ -396,8 +395,7 @@ def test_inc_get_quant_method_unquantized_routed_experts_with_model_prefix(

395 monkeypatch,

396 ) -> None:

397 """``model.``-prefixed extra_config entries trigger early exit for

399 - RoutedExperts layers, returning UnquantizedFusedMoEMethod.

400 - """

398 + RoutedExperts layers, returning UnquantizedFusedMoEMethod."""

399

400 class DummyUnquantizedFusedMoEMethod:

401 def __init__(self, moe_config) -> None:

@@ -420,8 +418,7 @@ def test_inc_get_quant_method_unquantized_routed_experts_with_model_prefix(

418

419 def test_inc_get_quant_method_unknown_layer_with_model_prefix_returns_none() -> None:

420 """``model.``-prefixed extra_config entries return None for unhandled

423 - layer types (non-Linear, non-MoE).

424 - """

421 + layer types (non-Linear, non-MoE)."""

422 config = make_config(extra_config={"model.unknown": {"bits": 16}})

423

424 method = config.get_quant_method(DummyLayer(), "unknown")

@@ -736,8 +733,7 @@ def test_wna16_cuda_low_bit_moe_routes_to_humming(monkeypatch, bits) -> None:

733 @pytest.mark.parametrize("bits", [4, 8])

734 def test_wna16_cuda_high_bit_skips_humming(monkeypatch, bits) -> None:

735 """4/8-bit int stays on the Marlin/GPTQ/AWQ path even on CUDA so a single

739 - model can mix high-bit Marlin and low-bit humming layers.

740 - """

736 + model can mix high-bit Marlin and low-bit humming layers."""

737 called = {"humming": False}

738

739 monkeypatch.setattr(current_platform, "is_cuda", lambda: True)

@@ -1668,8 +1664,7 @@ def test_inc_get_quant_method_unquantized_moe_returns_unquantized(

1664 monkeypatch,

1665 ) -> None:

1666 """Early-exit returns UnquantizedFusedMoEMethod for FusedMoE layers

1671 - when extra_config has bits >= 16.

1672 - """

1667 + when extra_config has bits >= 16."""

1668 config = make_config(extra_config={"layer": {"bits": 16}})

1669 layer = object.__new__(RoutedExperts)

1670 layer.moe_config = None # UnquantizedFusedMoEMethod accepts moe_config

tests/quantization/test_fp8.py

+1 -2

@@ -213,8 +213,7 @@ def test_deepseek_v41_vl_mapper_routes_linear_scales(

213 ):

214 """The VL wrapper must map ``.scale`` keys to the parameter the linear

215 quant method registers, as the text model does. A hardcoded

216 - ``weight_scale_inv`` raised KeyError for native MXFP8 checkpoints.

217 - """

216 + ``weight_scale_inv`` raised KeyError for native MXFP8 checkpoints."""

217 from vllm.models.deepseek_v41.nvidia import model as model_module

218 from vllm.models.deepseek_v41.nvidia import vl_model as vl_module

219

tests/quantization/test_gfx950_moe.py

+1 -2

@@ -160,8 +160,7 @@ def test_qwen38_tep8_auto_fallback_respects_explicit_backend(

160 @pytest.fixture

161 def mxfp4_oracle_config():

162 """Stub the config the oracle reads (``model_config.quantization_config``)

163 - so backend dispatch resolves without a real model / user override.

164 - """

163 + so backend dispatch resolves without a real model / user override."""

164 from unittest.mock import patch

165

166 with patch(

tests/quantization/test_humming_mxfp4_block_fp8.py

+2 -4

@@ -85,8 +85,7 @@ def test_block_fp8_activation_quant_config_is_block_quantized():

85 (FP8 dtype, [1, 128] block, not per-act-token). This is what makes the

86 DeepEP prepare/finalize step quantize activations to block FP8 *before* the

87 all-to-all dispatch instead of deferring to Humming -- and what makes

88 - HummingExpertsBase.expects_unquantized_inputs return False.

89 - """

88 + HummingExpertsBase.expects_unquantized_inputs return False."""

89 from vllm.model_executor.layers.quantization.utils.humming_utils import (

90 make_humming_moe_quant_config,

91 )

@@ -108,8 +107,7 @@ def test_block_fp8_activation_quant_config_is_block_quantized():

107 def test_default_activation_quant_config_defers_to_humming():

108 """Without a block activation shape the config stays per-token (the deferred

109 path): Humming quantizes internally, preserving pre-existing behavior for

111 - every non-block-FP8 scheme.

112 - """

110 + every non-block-FP8 scheme."""

111 from vllm.model_executor.layers.quantization.utils.humming_utils import (

112 make_humming_moe_quant_config,

113 )

tests/quantization/test_int8_moe_oracle.py

+1 -2

@@ -70,8 +70,7 @@ def _make_int8_moe_config(moe_backend: str = "auto") -> FusedMoEConfig:

70 )

71 def test_int8_dynamic_schemes_dispatch_to_triton(weight_key, activation_key):

72 """Both dynamic-activation INT8 MoE schemes (per-channel + per-tensor

73 - weights) select the Triton backend.

74 - """

73 + weights) select the Triton backend."""

74 config = _make_int8_moe_config()

75 backend, experts_cls = select_int8_moe_backend(

76 config, weight_key=weight_key, activation_key=activation_key

tests/quantization/test_modelopt.py

+4 -8

@@ -587,8 +587,7 @@ def test_modelopt_fp8_pb_wo_checkpoint_setup(monkeypatch, dist_init, workspace_i

587 def test_modelopt_nvfp4_config_dispatches_w4a4_method():

588 """``quant_method="NVFP4"`` (W4A4) resolves to a

589 ``(kNvfp4Static, kNvfp4Dynamic)`` QuantSpec under the generic

590 - ``ModelOptLinearMethod``.

591 - """

590 + ``ModelOptLinearMethod``."""

591 from vllm.model_executor.layers.linear import LinearBase

592

593 config = ModelOptNvFp4Config(

@@ -637,8 +636,7 @@ def test_modelopt_linear_method_builder_registry_override(monkeypatch):

636 """The bespoke-method escape hatch: a format registered in

637 ``LINEAR_METHOD_BUILDERS`` routes that algo to its own method instead of the

638 generic ``ModelOptLinearMethod``. This is how a format that cannot be a

640 - ``(weight, activation)`` key pair plugs into dispatch.

641 - """

639 + ``(weight, activation)`` key pair plugs into dispatch."""

640 from vllm.model_executor.layers.linear import LinearBase

641 from vllm.model_executor.layers.quantization import modelopt as m

642

@@ -669,8 +667,7 @@ def test_modelopt_linear_method_builder_registry_override(monkeypatch):

667 def test_modelopt_w4a16_respects_linear_backend(linear_backend, kernel_cls):

668 """W4A16 (`activation=None`) kernel selection honors ``--linear-backend``:

669 ``use_a16=True`` defaults to Marlin, but an explicit backend wins. The

672 - generic method routes this through ``select_linear_kernel``.

673 - """

670 + generic method routes this through ``select_linear_kernel``."""

671 from vllm.config.quantization import QuantSpec

672 from vllm.model_executor.layers.quantization.modelopt import (

673 RuntimeDtypes,

@@ -941,8 +938,7 @@ def test_modelopt_fp8_pb_wo_hides_output_padding(monkeypatch):

938 def test_modelopt_fp8_pb_wo_rejects_non_128_input():

939 """Input width must still be a multiple of 128 (same as #53132, which only

940 pads the output). A partial input block is refused loudly rather than

944 - silently loading wrong scales.

945 - """

941 + silently loading wrong scales."""

942 from vllm.model_executor.layers.quantization import modelopt as mo

943 from vllm.model_executor.layers.quantization.utils.quant_utils import (

944 kFp8Static128BlockSym,

tests/quantization/test_online_mxfp4.py

+1 -2

@@ -411,8 +411,7 @@ def test_online_mxfp4_moe_matches_quark(

411 ) -> tuple[torch.Tensor, torch.Tensor]:

412 """Place the source weights into buffers of the padded size, the

413 way the weight loader writes only the unpadded slice of a larger

414 - allocation.

415 - """

414 + allocation."""

415 w13 = torch.full(

416 (num_experts, 2 * intermediate_size, hidden_size),

417 padding_value,

tests/quantization/test_per_token_kv_cache.py

+1 -2

@@ -490,8 +490,7 @@ def test_per_token_head_negative_slot_skipped(qcfg: QuantConfig):

490 )

491 def test_process_weights_sets_placeholder_scales(kv_cache_dtype: str):

492 """Per-token-head should set _k_scale=1.0, _v_scale=1.0

493 - and delete checkpoint attrs.

494 - """

493 + and delete checkpoint attrs."""

494 from vllm.model_executor.layers.quantization.kv_cache import (

495 BaseKVCacheMethod,

496 )

tests/reasoning/test_kimi_k3_reasoning_parser.py

+4 -8

@@ -276,8 +276,7 @@ def _reference_is_reasoning_end(input_ids: list[int]) -> bool:

276 def test_is_reasoning_end_streaming_only_scans_the_step_window():

277 """The decode-step check must not re-derive the answer from the whole

278 sequence: an already-closed think block earlier in the sequence is the

279 - engine's business (it latches `reasoning_ended`), not this call's.

280 - """

279 + engine's business (it latches `reasoning_ended`), not this call's."""

280 parser = KimiK3ReasoningParser(DummyTokenizer())

281 prompt = [*CLOSE_IDS, *OPEN_IDS, 9, 9, 9]

282

@@ -287,8 +286,7 @@ def test_is_reasoning_end_streaming_only_scans_the_step_window():

286

287 def test_is_reasoning_end_streaming_detects_marker_across_step_boundary():

288 """A 3-token marker can be split over decode steps; the check carries the

290 - preceding len(marker)-1 tokens so the final token still completes it.

291 - """

289 + preceding len(marker)-1 tokens so the final token still completes it."""

290 parser = KimiK3ReasoningParser(DummyTokenizer())

291 history = [*OPEN_IDS, 5, *CLOSE_IDS[:2]]

292

@@ -299,8 +297,7 @@ def test_is_reasoning_end_streaming_detects_marker_across_step_boundary():

297 def test_is_reasoning_end_streaming_reopened_block_is_not_ended():

298 """Kimi K3 may close and immediately reopen the think channel inside one

299 window (speculative decoding lands several tokens per step). The newest

302 - marker wins, matching is_reasoning_end.

303 - """

300 + marker wins, matching is_reasoning_end."""

301 parser = KimiK3ReasoningParser(DummyTokenizer())

302 window = [*CLOSE_IDS, *OPEN_IDS]

303

@@ -327,8 +324,7 @@ def test_is_reasoning_end_streaming_thinking_disabled():

324 def test_reasoning_end_matches_reference_over_marker_dense_sequences(seed):

325 """Both the full-sequence check and the per-step check must agree with the

326 two-scan reference, on sequences where the markers (which share the suffix

330 - [2, 3]) collide and overlap constantly.

331 - """

327 + [2, 3]) collide and overlap constantly."""

328 import random

329

330 rnd = random.Random(seed)

tests/renderers/test_chat_utils_prompt_embeds.py

+3 -6

@@ -116,8 +116,7 @@ def test_prompt_embeds_keys_registered():

116

117 def test_ensure_placeholder_token_is_single_token_and_idempotent(tokenizer):

118 """Ensure the placeholder token is a single token and that multiple calls to

119 - "ensure" are idempotent, across all tokenizer families.

120 - """

119 + "ensure" are idempotent, across all tokenizer families."""

120 tid1 = _ensure_prompt_embeds_placeholder_token(tokenizer)

121 tid2 = _ensure_prompt_embeds_placeholder_token(tokenizer)

122 assert tid1 == tid2

@@ -470,8 +469,7 @@ def test_build_mixed_prompt_embeds(stream):

469 @pytest.mark.parametrize("role", ["user", "system"])

470 async def test_end_to_end_expand_and_build(tokenizer, parse_fn, role):

471 """Full renderer pipeline: parse -> chat template -> expand -> locate

473 - -> build mixed prompt, across tokenizers, roles, and sync/async.

474 - """

472 + -> build mixed prompt, across tokenizers, roles, and sync/async."""

473 tokenizer.chat_template = _SIMPLE_CHAT_TEMPLATE

474 tid = _ensure_prompt_embeds_placeholder_token(tokenizer)

475

@@ -526,8 +524,7 @@ async def test_end_to_end_expand_and_build(tokenizer, parse_fn, role):

524 @pytest.mark.asyncio

525 async def test_end_to_end_multi_message_conversation(tokenizer, parse_fn):

526 """Full pipeline with prompt_embeds spread across system + user messages,

529 - verifying ordering and positioning in the final token stream.

530 - """

527 + verifying ordering and positioning in the final token stream."""

528 tokenizer.chat_template = _SIMPLE_CHAT_TEMPLATE

529 tid = _ensure_prompt_embeds_placeholder_token(tokenizer)

530

tests/renderers/test_gemma4_chat_template.py

+9 -18

@@ -33,8 +33,7 @@ def _render(template, messages, **kwargs):

33 class TestGemma4ChatTemplate:

34 def test_basic_multiturn_thinking_disabled(self, gemma4_template):

35 """With enable_thinking=False (default), generation prompt ends with

36 - an empty thought channel to suppress thinking.

37 - """

36 + an empty thought channel to suppress thinking."""

37 messages = [

38 {"role": "user", "content": "Hello"},

39 {"role": "assistant", "content": "Hi there!"},

@@ -50,8 +49,7 @@ class TestGemma4ChatTemplate:

49

50 def test_basic_multiturn_thinking_enabled(self, gemma4_template):

51 """With enable_thinking=True, generation prompt ends with model

53 - turn opener (no thought suppression).

54 - """

52 + turn opener (no thought suppression)."""

53 messages = [

54 {"role": "user", "content": "Hello"},

55 {"role": "assistant", "content": "Hi there!"},

@@ -146,8 +144,7 @@ class TestGemma4ChatTemplate:

144

145 def test_tool_responses_openai_style(self, gemma4_template):

146 """role='tool' messages are formatted as <|tool_response> blocks

149 - with content dumped as-is.

150 - """

147 + with content dumped as-is."""

148 messages = [

149 {"role": "user", "content": "Weather?"},

150 {

@@ -205,8 +202,7 @@ class TestGemma4ChatTemplate:

202

203 def test_generation_prompt_not_after_tool_response(self, gemma4_template):

204 """add_generation_prompt=True should NOT add <|turn>model when the

208 - last message type was tool_response (the model turn continues).

209 - """

205 + last message type was tool_response (the model turn continues)."""

206 messages = [

207 {"role": "user", "content": "Weather?"},

208 {

@@ -233,8 +229,7 @@ class TestGemma4ChatTemplate:

229

230 def test_reasoning_in_tool_chains(self, gemma4_template):

231 """Reasoning field on assistant with tool_calls after last user

236 - message emits <|channel>thought\\n...<channel|>.

237 - """

232 + message emits <|channel>thought\\n...<channel|>."""

233 messages = [

234 {"role": "user", "content": "Calculate something"},

235 {

@@ -280,8 +275,7 @@ class TestGemma4ChatTemplate:

275

276 def test_strip_thinking_in_model_content(self, gemma4_template):

277 """<|channel>...<channel|> in model content is stripped by the

283 - strip_thinking macro.

284 - """

278 + strip_thinking macro."""

279 messages = [

280 {"role": "user", "content": "Hi"},

281 {

@@ -295,8 +289,7 @@ class TestGemma4ChatTemplate:

289

290 def test_multi_turn_tool_chain(self, gemma4_template):

291 """assistant->tool->assistant->tool produces exactly one

298 - <|turn>model (later assistants continue the same turn).

299 - """

292 + <|turn>model (later assistants continue the same turn)."""

293 messages = [

294 {"role": "user", "content": "Do two things"},

295 {

@@ -353,8 +346,7 @@ class TestGemma4ChatTemplate:

346

347 def test_tool_response_with_multimodal_content(self, gemma4_template):

348 """Multimodal placeholders in tool messages are emitted after the

356 - tool_response block.

357 - """

349 + tool_response block."""

350 messages = [

351 {"role": "user", "content": "Download the image and describe it."},

352 {

@@ -388,8 +380,7 @@ class TestGemma4ChatTemplate:

380

381 def test_tool_response_with_all_modalities(self, gemma4_template):

382 """All multimodal types (image, audio, video) in a single tool

391 - response are rendered.

392 - """

383 + response are rendered."""

384 messages = [

385 {"role": "user", "content": "Process media"},

386 {

tests/renderers/test_hf.py

+3 -6

@@ -308,8 +308,7 @@ def test_resolve_chat_template_kwargs(sample_json_schema, model, expected_kwargs

308

309 def test_resolve_chat_template_resolves_name():

310 """When chat_template is a name, resolve_chat_template should return

311 - the actual Jinja content so that kwargs detection works correctly.

312 - """

311 + the actual Jinja content so that kwargs detection works correctly."""

312 from unittest.mock import MagicMock

313

314 jinja_content = "{{ messages }}{% if tools %}{{ tools }}{% endif %}"

@@ -331,8 +330,7 @@ def test_resolve_chat_template_resolves_name():

330

331 def test_resolve_chat_template_kwargs_with_template_name():

332 """Ensures template kwargs are not silently dropped when chat_template

334 - was originally a template name that has been resolved to Jinja content.

335 - """

333 + was originally a template name that has been resolved to Jinja content."""

334 from unittest.mock import MagicMock

335

336 jinja_content = (

@@ -934,8 +932,7 @@ EFFORT_BREAKING_TEMPLATE = (

932

933 class TestApplyChatTemplateEffortTolerant:

934 """Chat templates that reject unsupported reasoning_effort values should

937 - surface a 400-style client error instead of crashing the request.

938 - """

935 + surface a 400-style client error instead of crashing the request."""

936

937 @pytest.fixture

938 def model_config(self):

tests/renderers/test_process_multi_modal_uuids.py

+1 -2

@@ -49,8 +49,7 @@ def _build_text_only_renderer() -> HfRenderer:

49

50 def test_text_only_model_mm_data_maps_to_bad_request():

51 """Sending multimodal data to a text-only model is a client mistake, so it

52 - must surface as a ValueError and reach the client as HTTP 400, not 500.

53 - """

52 + must surface as a ValueError and reach the client as HTTP 400, not 500."""

53 renderer = _build_text_only_renderer()

54

55 with pytest.raises(ValueError, match="text-only") as exc_info:

tests/renderers/test_prompt_embeds_payload_validation.py

+1 -2

@@ -127,8 +127,7 @@ def test_a_long_reason_is_truncated_and_says_how_much_was_left_out(model_config)

127

128 def test_a_short_reason_is_passed_through_whole(model_config):

129 """A reason that fits is not annotated -- the count only appears when it is

130 - telling the reader something.

131 - """

130 + telling the reader something."""

131 reason = "invalid magic number; corrupt file?"

132 assert len(reason) <= _MAX_EMBED_ERROR_REASON_CHARS

133

tests/renderers/test_sparse_tensor_concurrent_race.py

+1 -2

@@ -218,8 +218,7 @@ class TestCrossLoaderLockSharing:

218

219 def test_prompt_and_image_share_lock(self, model_config):

220 """Prompt and image loaders cannot run their sparse guards

221 - concurrently.

222 - """

221 + concurrently."""

222 valid_prompt = _encode_tensor(_create_valid_dense_tensor())

223 valid_image = _encode_tensor(_create_valid_dense_tensor(hidden_size=10))

224

tests/renderers/test_sparse_tensor_validation.py

+7 -14

@@ -157,8 +157,7 @@ class TestPromptEmbedsValidation:

157

158 def test_hidden_size_mismatch_rejected(self, model_config):

159 """Tensors whose trailing dim doesn't match the model's hidden_size

160 - must be rejected at parse time.

161 - """

160 + must be rejected at parse time."""

161 # opt-125m has hidden_size=768, passing 512 triggers the check.

162 wrong_hidden = torch.randn(10, 512, dtype=torch.float32)

163 encoded = _encode_tensor(wrong_hidden)

@@ -169,8 +168,7 @@ class TestPromptEmbedsValidation:

168 def test_float_dtype_mismatch_cast_to_model_dtype(self, model_config):

169 """Tensors whose dtype doesn't match the model's dtype but are still

170 floating-point are cast, since API clients generally can't know the

172 - server's `--dtype` setting ahead of time.

173 - """

171 + server's `--dtype` setting ahead of time."""

172 # Fixture pins model dtype to float32, upload a bfloat16 tensor.

173 mismatched_float = torch.randn(10, 768, dtype=torch.bfloat16)

174 encoded = _encode_tensor(mismatched_float)

@@ -183,8 +181,7 @@ class TestPromptEmbedsValidation:

181 def test_non_float_dtype_rejected(self, model_config):

182 """Non-floating-point dtypes cannot be safely cast for embeddings

183 (e.g. integer tensors almost certainly indicate caller confusion),

186 - so they are rejected at parse time.

187 - """

184 + so they are rejected at parse time."""

185 non_float = torch.randint(0, 100, (10, 768), dtype=torch.int32)

186 encoded = _encode_tensor(non_float)

187

@@ -193,8 +190,7 @@ class TestPromptEmbedsValidation:

190

191 def test_non_2d_tensor_rejected(self, model_config):

192 """Tensors that aren't 2D (even after squeezing a leading dim)

196 - must be rejected with a clear error.

197 - """

193 + must be rejected with a clear error."""

194 # A 1D tensor cannot be interpreted as (num_tokens, hidden_size).

195 bad = torch.randn(768, dtype=torch.float32)

196 encoded = _encode_tensor(bad)

@@ -204,8 +200,7 @@ class TestPromptEmbedsValidation:

200

201 def test_non_tensor_payload_rejected(self, model_config):

202 """Deserializing to a non-Tensor object must raise a clear error

207 - instead of propagating an AssertionError.

208 - """

203 + instead of propagating an AssertionError."""

204 # `torch.save` will serialize a plain dict; `weights_only=True` allows

205 # loading built-in containers, so this exercises the isinstance check.

206 buffer = io.BytesIO()

@@ -463,8 +458,7 @@ class TestEmbeddingDecodeSizeLimit:

458 def test_default_limit_would_reject_the_real_payload(self):

459 """The shipped default has to bound the actual attack, not just the

460 shrunken shape used above. Checked arithmetically so no test ever asks

466 - the allocator for 4 TiB.

467 - """

461 + the allocator for 4 TiB."""

462 dense_bytes = BOMB_SHAPE[0] * BOMB_SHAPE[1] * 4

463 assert dense_bytes > envs.VLLM_MAX_EMBED_DECODE_BYTES > 0

464

@@ -522,8 +516,7 @@ class TestEmbeddingDecodeSizeLimit:

516 @pytest.mark.parametrize("io_cls,parameter", EMBEDDING_MEDIA_IO)

517 def test_media_io_rejects_non_tensor_payload(self, io_cls, parameter):

518 """These three had no `isinstance` check, so a non-tensor payload

525 - surfaced as an `AttributeError` (500) rather than a 4xx.

526 - """

519 + surfaced as an `AttributeError` (500) rather than a 4xx."""

520 buffer = io.BytesIO()

521 torch.save({"not": "a tensor"}, buffer)

522 encoded = base64.b64encode(buffer.getvalue()).decode("utf-8")

tests/renderers/test_token_offsets.py

+5 -10

@@ -26,8 +26,7 @@ def fast_tokenizer():

26 def _make_base_renderer_with(tokenizer):

27 """Build a minimal BaseRenderer subclass that exposes the tokenizer so we

28 can call ``_tokenize_prompt`` directly. BaseRenderer is abstract because of

29 - ``render_messages``; we just need a stub.

30 - """

29 + ``render_messages``; we just need a stub."""

30 from vllm.renderers.base import BaseRenderer

31

32 class _StubRenderer(BaseRenderer):

@@ -92,8 +91,7 @@ class TestTokenizePromptOffsets:

91 def test_base_renderer_without_override_yields_no_offsets(self, fast_tokenizer):

92 """A renderer that does not override ``_can_produce_offsets`` never

93 emits offsets, even with a fast tokenizer and the flag set. This locks

95 - in the base-default-False / subclass-override design.

96 - """

94 + in the base-default-False / subclass-override design."""

95 from vllm.renderers.base import BaseRenderer

96

97 class _BareRenderer(BaseRenderer):

@@ -127,8 +125,7 @@ class TestTokenizePromptOffsets:

125

126 def test_slow_tokenizer_with_flag_no_offsets(self, fast_tokenizer):

127 """Force is_fast=False to simulate a Slow tokenizer: the flag is set

130 - but offsets must not be returned because it cannot produce them.

131 - """

128 + but offsets must not be returned because it cannot produce them."""

129 from unittest.mock import PropertyMock, patch

130

131 renderer = _make_base_renderer_with(fast_tokenizer)

@@ -147,8 +144,7 @@ class TestTokenizePromptOffsets:

144 @pytest.mark.parametrize("mm_key", ["multi_modal_data", "multi_modal_uuids"])

145 def test_multimodal_with_flag_no_offsets(self, fast_tokenizer, mm_key):

146 """Offsets index the text prompt, which is meaningless once multimodal

150 - data is interleaved, so they are suppressed when MM inputs are present.

151 - """

147 + data is interleaved, so they are suppressed when MM inputs are present."""

148 renderer = _make_base_renderer_with(fast_tokenizer)

149 params = TokenizeParams(max_total_tokens=None, return_token_offsets=True)

150 prompt = {"prompt": "Hello.", mm_key: {"image": ["x"]}}

@@ -160,8 +156,7 @@ class TestTokenizePromptOffsets:

156 @pytest.mark.asyncio

157 async def test_tokenize_prompt_async_returns_offsets(self, fast_tokenizer):

158 """The async path offloads the sync tokenizer; it must yield the same

163 - offsets as the sync path.

164 - """

159 + offsets as the sync path."""

160 renderer = _make_base_renderer_with(fast_tokenizer)

161 params = TokenizeParams(max_total_tokens=None, return_token_offsets=True)

162

tests/test_cmake_utils.py

+1 -2

@@ -40,8 +40,7 @@ def test_extract_archs_prefers_sass_target_over_corrupted_virtual_arch(

40 ):

41 """Torch's autodetection can emit a bogus arch=compute_* half (e.g.

42 capability 12.1 corrupted to arch=compute_20,code=sm_121); the SASS

43 - target must win, while PTX-only entries keep the virtual arch.

44 - """

43 + target must win, while PTX-only entries keep the virtual arch."""

44 repo_root = Path(__file__).parents[1]

45 script = tmp_path / "test_extract_archs.cmake"

46 script.write_text(

tests/test_config.py

+10 -20

@@ -441,8 +441,7 @@ def test_no_hisparse_connector_keeps_attention_config_unset(monkeypatch):

441

442 def test_rocm_keeps_compiled_deepseek_defaults(monkeypatch):

443 """ROCm keeps the DSA models (DeepSeek V3.2/V4, GLM-5.2) on their compiled

444 - MRV1 paths and off breakable cudagraphs by default.

445 - """

444 + MRV1 paths and off breakable cudagraphs by default."""

445 from vllm.config.vllm import (

446 ROCM_DEFAULT_MRV1_ARCHITECTURES,

447 default_breakable_cudagraph_architectures,

@@ -657,8 +656,7 @@ def test_v2_model_runner_supports_extract_hidden_states():

656

657 def test_dflash2_draft_forces_v2_model_runner():

658 """A DFlash2 draft must reach the V2 speculator, the only one that runs its

660 - candidate selector; on V1 it would draft as DFlash1 without raising.

661 - """

659 + candidate selector; on V1 it would draft as DFlash1 without raising."""

660

661 def config(method, architectures):

662 return SimpleNamespace(

@@ -771,8 +769,7 @@ def test_late_piecewise_restrictions_without_compilation(monkeypatch, engine_kwa

769

770 def test_resolve_cudagraph_mode_skips_mamba_block_check_while_profiling():

771 """Cudagraph memory profiling uses a minimal KV cache, so the Mamba

774 - block-count guard must only fire for the real cache sizing.

775 - """

772 + block-count guard must only fire for the real cache sizing."""

773 kv_cache_config = SimpleNamespace(has_mamba_layers=True, num_blocks=4)

774

775 compilation_config = CompilationConfig(

@@ -1097,8 +1094,7 @@ def test_v1_model_runner_rejects_v2_only_features():

1094

1095 def test_batch_sharded_sampling_rejects_return_sampling_mask():

1096 """The batch-sharded gather drops sampling masks, so the combination must

1100 - fail loudly instead of returning ``sampling_mask=None``.

1101 - """

1097 + fail loudly instead of returning ``sampling_mask=None``."""

1098 config = SimpleNamespace(

1099 parallel_config=SimpleNamespace(

1100 enable_batch_sharded_sampling=True, tensor_parallel_size=2

@@ -2195,8 +2191,7 @@ class MockConfig:

2191 @patch("vllm.transformers_utils.runai_utils.ObjectStorageModel.pull_files")

2192 def test_s3_url_model_tokenizer_paths(mock_pull_files, s3_url):

2193 """Test that S3 URLs create deterministic local directories for model and

2198 - tokenizer.

2199 - """

2194 + tokenizer."""

2195 # Mock pull_files to avoid actually downloading files during tests

2196 mock_pull_files.return_value = None

2197

@@ -2294,8 +2289,7 @@ def test_s3_url_different_models_create_different_directories(mock_pull_files):

2289 @patch("vllm.transformers_utils.runai_utils.ObjectStorageModel.pull_files")

2290 def test_s3_url_different_model_and_tokenizer(mock_pull_files):

2291 """Test that when model and tokenizer are different cloud URIs,

2297 - pull_files receives the correct URI for each.

2298 - """

2292 + pull_files receives the correct URI for each."""

2293 mock_pull_files.return_value = None

2294

2295 model_url = "s3://bucket/model/"

@@ -2840,8 +2834,7 @@ def test_vllm_config_explicit_overrides():

2834

2835 def test_fusion_pass_op_priority():

2836 """This test checks that custom op enablement & IR op priority

2843 - correctly control default fusions

2844 - """

2837 + correctly control default fusions"""

2838 # Default config, O2, rms_norm+quant fusion disabled

2839 cfg1 = VllmConfig()

2840 assert not cfg1.compilation_config.pass_config.fuse_norm_quant

@@ -2925,8 +2918,7 @@ def test_needs_dp_coordination(

2918

2919 def test_fault_tolerance_requires_single_api_server():

2920 """Fault tolerance assumes one AsyncMPClient manages all engines, so it

2928 - is incompatible with API server scale-out (_api_process_count > 1).

2929 - """

2921 + is incompatible with API server scale-out (_api_process_count > 1)."""

2922 with pytest.raises(ValueError, match="single API server"):

2923 ParallelConfig(enable_fault_tolerance=True, _api_process_count=2)

2924

@@ -2936,8 +2928,7 @@ def test_fault_tolerance_requires_single_api_server():

2928

2929 def test_renderer_num_workers_with_mm_cache():

2930 """Disallow renderer_num_workers > 1 with the mm processor cache only for

2939 - pooling models, whose preprocessing runs on the renderer workers.

2940 - """

2931 + pooling models, whose preprocessing runs on the renderer workers."""

2932 mm_model = "Qwen/Qwen2-VL-2B-Instruct"

2933

2934 # Should raise: pooling + multi-worker + cache enabled (default cache_gb=4)

@@ -3263,8 +3254,7 @@ def test_watermarking_forces_model_runner_v2(monkeypatch):

3254 def test_mtp_draft_uses_model_weights_not_local_cache(mock_model_config_cls):

3255 """Regression test: MTP + runai_streamer should use model_weights (original

3256 S3 URL) for the draft model, not model (local cache dir set by

3266 - pull_runai_model_from_obj_storage).

3267 - """

3257 + pull_runai_model_from_obj_storage)."""

3258 from unittest.mock import MagicMock

3259

3260 s3_url = "s3://my-bucket/Qwen3-35B-A3B-FP8"

tests/test_logger.py

+8 -16

@@ -51,8 +51,7 @@ def test_trace_function_call():

51 def test_default_vllm_root_logger_configuration(monkeypatch):

52 """This test presumes that VLLM_CONFIGURE_LOGGING (default: True) and

53 VLLM_LOGGING_CONFIG_PATH (default: None) are not configured and default

54 - behavior is activated.

55 - """

54 + behavior is activated."""

55 monkeypatch.setenv("VLLM_LOGGING_COLOR", "0")

56 _configure_vllm_root_logger()

57

@@ -76,8 +75,7 @@ def test_default_vllm_root_logger_configuration(monkeypatch):

75 def test_descendent_loggers_depend_on_and_propagate_logs_to_root_logger(monkeypatch):

76 """This test presumes that VLLM_CONFIGURE_LOGGING (default: True) and

77 VLLM_LOGGING_CONFIG_PATH (default: None) are not configured and default

79 - behavior is activated.

80 - """

78 + behavior is activated."""

79 monkeypatch.setenv("VLLM_CONFIGURE_LOGGING", "1")

80 monkeypatch.delenv("VLLM_LOGGING_CONFIG_PATH", raising=False)

81

@@ -107,8 +105,7 @@ def test_descendent_loggers_depend_on_and_propagate_logs_to_root_logger(monkeypa

105 def test_logger_configuring_can_be_disabled(monkeypatch):

106 """This test calls _configure_vllm_root_logger again to test custom logging

107 config behavior, however mocks are used to ensure no changes in behavior or

110 - configuration occur.

111 - """

108 + configuration occur."""

109 monkeypatch.setenv("VLLM_CONFIGURE_LOGGING", "0")

110 monkeypatch.delenv("VLLM_LOGGING_CONFIG_PATH", raising=False)

111

@@ -120,8 +117,7 @@ def test_logger_configuring_can_be_disabled(monkeypatch):

117 def test_an_error_is_raised_when_custom_logging_config_file_does_not_exist(monkeypatch):

118 """This test calls _configure_vllm_root_logger again to test custom logging

119 config behavior, however it fails before any change in behavior or

123 - configuration occurs.

124 - """

120 + configuration occurs."""

121 monkeypatch.setenv("VLLM_CONFIGURE_LOGGING", "1")

122 monkeypatch.setenv(

123 "VLLM_LOGGING_CONFIG_PATH",

@@ -137,8 +133,7 @@ def test_an_error_is_raised_when_custom_logging_config_file_does_not_exist(monke

133 def test_an_error_is_raised_when_custom_logging_config_is_invalid_json(monkeypatch):

134 """This test calls _configure_vllm_root_logger again to test custom logging

135 config behavior, however it fails before any change in behavior or

140 - configuration occurs.

141 - """

136 + configuration occurs."""

137 monkeypatch.setenv("VLLM_CONFIGURE_LOGGING", "1")

138

139 with NamedTemporaryFile(encoding="utf-8", mode="w") as logging_config_file:

@@ -165,8 +160,7 @@ def test_an_error_is_raised_when_custom_logging_config_is_unexpected_json(

160 ):

161 """This test calls _configure_vllm_root_logger again to test custom logging

162 config behavior, however it fails before any change in behavior or

168 - configuration occurs.

169 - """

163 + configuration occurs."""

164 monkeypatch.setenv("VLLM_CONFIGURE_LOGGING", "1")

165

166 with NamedTemporaryFile(encoding="utf-8", mode="w") as logging_config_file:

@@ -182,8 +176,7 @@ def test_an_error_is_raised_when_custom_logging_config_is_unexpected_json(

176 def test_custom_logging_config_is_parsed_and_used_when_provided(monkeypatch):

177 """This test calls _configure_vllm_root_logger again to test custom logging

178 config behavior, however mocks are used to ensure no changes in behavior or

185 - configuration occur.

186 - """

179 + configuration occur."""

180 monkeypatch.setenv("VLLM_CONFIGURE_LOGGING", "1")

181

182 valid_logging_config = {

@@ -207,8 +200,7 @@ def test_custom_logging_config_is_parsed_and_used_when_provided(monkeypatch):

200 def test_custom_logging_config_causes_an_error_if_configure_logging_is_off(monkeypatch):

201 """This test calls _configure_vllm_root_logger again to test custom logging

202 config behavior, however mocks are used to ensure no changes in behavior or

210 - configuration occur.

211 - """

203 + configuration occur."""

204 monkeypatch.setenv("VLLM_CONFIGURE_LOGGING", "0")

205

206 valid_logging_config = {

tests/test_ray_env.py

+2 -4

@@ -15,8 +15,7 @@ from vllm.v1.executor.ray_utils import WORKER_SPECIFIC_ENV_VARS

15

16 class TestDefaultPrefixes:

17 """Built-in prefixes (VLLM_, LMCACHE_, NCCL_, UCX_, HF_, HUGGING_FACE_)

18 - should be forwarded without any extra configuration.

19 - """

18 + should be forwarded without any extra configuration."""

19

20 @patch.dict(os.environ, {"LMCACHE_LOCAL_CPU": "True"}, clear=False)

21 def test_lmcache_prefix(self):

@@ -54,8 +53,7 @@ class TestDefaultExtraVars:

53

54 def test_pythonhashseed_in_result(self):

55 """PYTHONHASHSEED should always be in the result set (as a name to

57 - copy) regardless of whether it is actually set in os.environ.

58 - """

56 + copy) regardless of whether it is actually set in os.environ."""

57 result = get_env_vars_to_copy()

58 assert "PYTHONHASHSEED" in result

59

tests/tokenizers_/test_deepseek_v4.py

+1 -2

@@ -452,8 +452,7 @@ def test_deepseek_v4_image_blocks_become_placeholders():

452

453 def test_deepseek_v4_image_sentinel_ids_match_tokenizer():

454 """The borrowed sentinel ids must line up with the reserved

455 - ``<|place_holder_mm_span_XXXX|>`` tokens in the tokenizer.

456 - """

455 + ``<|place_holder_mm_span_XXXX|>`` tokens in the tokenizer."""

456 from vllm.models.deepseek_v4.common.mm_preprocess import (

457 IMAGE_SENTINEL_BASE_ID,

458 IMAGE_SENTINEL_TOKEN_NAMES,

tests/tokenizers_/test_hf.py

+1 -2

@@ -51,8 +51,7 @@ def _check_consistency(target: TokenizerLike, expected: TokenizerLike):

51 def test_thread_pool_tokenizer_pickle(model_id: str):

52 """Regression test for issue #45433: the thread-pool tokenizer wrapper

53 reconstructs through maybe_make_thread_pool on unpickling, which used to

54 - fall off the end and return None.

55 - """

54 + fall off the end and return None."""

55 reference_tokenizer = AutoTokenizer.from_pretrained(model_id)

56

57 pooled_tokenizer = maybe_make_thread_pool(deepcopy(reference_tokenizer))

tests/tool_parsers/test_apertus_tool_parser.py

+4 -8

@@ -341,8 +341,7 @@ class TestStreamingExtraction:

341

342 def test_mtp_streaming_massive_chunk(self, parser, mock_request):

343 """Simulates MTP predicting text, tool calls,

344 - and trailing text all in a single chunk.

345 - """

344 + and trailing text all in a single chunk."""

345 chunks = [

346 "Sure! "

347 '<|tools_prefix|>[{"get_weather": {"location": "London"}}]<|tools_suffix|>'

@@ -375,8 +374,7 @@ class TestStreamingExtraction:

374

375 def test_mtp_streaming_skip_and_catch_up(self, parser, mock_request):

376 """Simulates MTP chunks that jump over entire tools

378 - (e.g., from middle of tool 1 to middle of tool 3).

379 - """

377 + (e.g., from middle of tool 1 to middle of tool 3)."""

378 chunks = [

379 '<|tools_prefix|>[{"t1": {"a": 1}',

380 '}, {"t2": {"b": 2}}, {"t3": {"c": 3',

@@ -395,8 +393,7 @@ class TestStreamingExtraction:

393

394 def test_vllm_streaming_character_by_character(self, parser, mock_request):

395 """Simulates worst-case vLLM fragmentation where

398 - chunks arrive character-by-character.

399 - """

396 + chunks arrive character-by-character."""

397 text = (

398 'Hi <|tools_prefix|>[{"get_weather": '

399 '{"location": "London"}}]<|tools_suffix|> '

@@ -414,8 +411,7 @@ class TestStreamingExtraction:

411

412 def test_vllm_streaming_empty_deltas(self, parser, mock_request):

413 """Simulates vLLM stream producing empty string chunks

417 - (e.g., hidden tokens or artifacts).

418 - """

414 + (e.g., hidden tokens or artifacts)."""

415 chunks = [

416 "Wait",

417 "",

tests/tool_parsers/test_deepseekv32_tool_parser.py

+10 -20

@@ -173,8 +173,7 @@ class TestExtractToolCalls:

173

174 def test_string_attr_true_coerced_by_schema(self):

175 """string="true" delivers a string, but the engine's schema-aware

176 - type fixer coerces it to the schema type (integer).

177 - """

176 + type fixer coerces it to the schema type (integer)."""

177 tool = ChatCompletionToolsParam(

178 function=FunctionDefinition(

179 name="score",

@@ -288,8 +287,7 @@ class TestExtractToolCalls:

287

288 def test_string_attr_true_coerced_by_composed_schema(self):

289 """string="true" delivers a JSON string, but the engine's schema-aware

291 - type fixer coerces it to the composed schema type (object).

292 - """

290 + type fixer coerces it to the composed schema type (object)."""

291 tool = ChatCompletionToolsParam(

292 function=FunctionDefinition(

293 name="set_timer",

@@ -323,8 +321,7 @@ class TestExtractToolCalls:

321

322 def test_arguments_wrapper_repaired(self):

323 """A single 'arguments' wrapper parameter must be unwrapped when it

326 - is not part of the tool schema and the inner object matches schema fields.

327 - """

324 + is not part of the tool schema and the inner object matches schema fields."""

325 tool = ChatCompletionToolsParam(

326 function=FunctionDefinition(

327 name="get_weather",

@@ -495,8 +492,7 @@ class TestExtractToolCalls:

492

493 def test_null_coerced_back_to_string_by_schema(self):

494 """string="false" with 'null' is json-parsed to None, but the

498 - engine's schema fixer coerces it back to "null" for string schemas.

499 - """

495 + engine's schema fixer coerces it back to "null" for string schemas."""

496 tool = ChatCompletionToolsParam(

497 function=FunctionDefinition(

498 name="echo",

@@ -659,8 +655,7 @@ class TestExtractToolCallsStreaming:

655

656 def test_string_attr_true_coerced_by_schema_streaming(self):

657 """Streaming: string='true' delivers a string but the engine's

662 - schema fixer coerces it to the schema type (integer).

663 - """

658 + schema fixer coerces it to the schema type (integer)."""

659 tool = ChatCompletionToolsParam(

660 function=FunctionDefinition(

661 name="score",

@@ -976,8 +971,7 @@ class TestExtractToolCallsStreaming:

971

972 def test_no_marker_leak_chunked(self, parser):

973 """Chunked streaming must NOT leak DSML start-marker fragments

979 - as content (GitHub #40801).

980 - """

974 + as content (GitHub #40801)."""

975 full_text = build_tool_call("fn", {"k": "v"})

976 deltas = self._stream_chunked(parser, full_text, chunk_size=5)

977 content = "".join(d.content for d in deltas if d.content is not None)

@@ -987,8 +981,7 @@ class TestExtractToolCallsStreaming:

981

982 def test_no_marker_leak_with_prefix_chunked(self, parser):

983 """Content before a tool call must not include start-marker

990 - fragments when chunked (GitHub #40801).

991 - """

984 + fragments when chunked (GitHub #40801)."""

985 full_text = "Hello!" + build_tool_call("fn", {"a": "b"})

986 deltas = self._stream_chunked(parser, full_text, chunk_size=5)

987 content = "".join(d.content for d in deltas if d.content is not None)

@@ -1000,8 +993,7 @@ class TestExtractToolCallsStreaming:

993

994 def test_no_marker_leak_char_by_char(self, parser):

995 """Character-by-character streaming must not leak marker

1003 - fragments (GitHub #40801).

1004 - """

996 + fragments (GitHub #40801)."""

997 full_text = build_tool_call("fn", {"k": "v"})

998 deltas = self._stream_chunked(parser, full_text, chunk_size=1)

999 content = "".join(d.content for d in deltas if d.content is not None)

@@ -1011,8 +1003,7 @@ class TestExtractToolCallsStreaming:

1003

1004 def test_no_marker_leak_all_split_points(self, parser):

1005 """Start token split at every possible boundary must not

1014 - leak (GitHub #40801).

1015 - """

1006 + leak (GitHub #40801)."""

1007 for chunk_size in range(1, len(FC_START) + 2):

1008 p = make_parser()

1009 full_text = build_tool_call("fn", {"k": "v"})

@@ -1024,8 +1015,7 @@ class TestExtractToolCallsStreaming:

1015

1016 def test_false_partial_marker_emitted(self, parser):

1017 """Text ending with a prefix of the start token that turns out

1027 - NOT to be a marker must still be emitted as content.

1028 - """

1018 + NOT to be a marker must still be emitted as content."""

1019 full_text = "<｜DSM some regular text"

1020 deltas = self._stream_chunked(parser, full_text, chunk_size=3)

1021 content = "".join(d.content for d in deltas if d.content is not None)

tests/tool_parsers/test_gemma4_tool_parser.py

+1 -2

@@ -124,8 +124,7 @@ def mock_request():

124

125 class TestParseGemma4Args:

126 """Values are returned as strings; type coercion to proper JSON types

127 - happens at the engine layer.

128 - """

127 + happens at the engine layer."""

128

129 def test_empty_string(self):

130 assert _parse_gemma4_args("") == {}

tests/tool_parsers/test_gigachat3_tool_parser.py

+1 -2

@@ -18,8 +18,7 @@ from vllm.tool_parsers import ToolParser, ToolParserManager

18 @pytest.fixture(scope="function")

19 def default_tokenizer() -> TokenizerLike:

20 """Override module-scoped default_tokenizer because gigachat tests

21 - mutate the tokenizer via ``add_tokens``.

22 - """

21 + mutate the tokenizer via ``add_tokens``."""

22 return AutoTokenizer.from_pretrained("openai-community/gpt2")

23

24

tests/tool_parsers/test_kimi_k3_named_tool_choice.py

+1 -2

@@ -1,8 +1,7 @@

1 # SPDX-License-Identifier: Apache-2.0

2 # SPDX-FileCopyrightText: Copyright contributors to the vLLM project

3 """Named tool choice for Kimi K3: allowed when the XTML structural tag is

4 -attached (strict tool calling), rejected otherwise.

5 -"""

4 +attached (strict tool calling), rejected otherwise."""

5

6 import pytest

7

tests/tool_parsers/test_lfm2_tool_parser.py

+9 -18

@@ -416,8 +416,7 @@ def test_tool_call(

416 def test_empty_tool_call_block(body: str, lfm2_tokenizer: TokenizerLike):

417 """An empty block ([] or [ ]) means the model called no tools; it must

418 not report tools_called=True with zero tool calls (the invariant

419 - run_tool_extraction asserts).

420 - """

419 + run_tool_extraction asserts)."""

420 cls = ToolParserManager.get_tool_parser("lfm2")

421 model_output = f"{TOOL_CALL_START}[{body}]{TOOL_CALL_END}"

422

@@ -437,8 +436,7 @@ def test_whitespace_after_start_token(lfm2_tokenizer: TokenizerLike):

436 not break parsing. The streaming path used to feed the indented text to

437 ast.parse verbatim, so every completion candidate raised

438 IndentationError and the call was silently dropped (non-streaming

440 - stripped it and succeeded).

441 - """

439 + stripped it and succeeded)."""

440 cls = ToolParserManager.get_tool_parser("lfm2")

441 model_output = f"{TOOL_CALL_START} [{SIMPLE_FUNCTION_OUTPUT}]{TOOL_CALL_END}"

442

@@ -478,8 +476,7 @@ def test_streaming_full_block_and_trailing_in_single_delta(

476 ):

477 """The entire <|tool_call_start|>[...]<|tool_call_end|> block plus

478 trailing assistant text arrive in one delta. Trailing content must

481 - still be emitted — not silently dropped.

482 - """

479 + still be emitted — not silently dropped."""

480 tool_parser: ToolParser = ToolParserManager.get_tool_parser("lfm2")(lfm2_tokenizer)

481 full_text = f"{TOOL_CALL_START}[{SIMPLE_FUNCTION_OUTPUT}]{TOOL_CALL_END}\nDone."

482

@@ -494,8 +491,7 @@ def test_streaming_leading_content_and_full_block_in_single_delta(

491 lfm2_tokenizer: TokenizerLike,

492 ):

493 """Leading assistant text plus the entire tool block arrive in one

497 - delta. Leading content must be emitted — not silently dropped.

498 - """

494 + delta. Leading content must be emitted — not silently dropped."""

495 tool_parser: ToolParser = ToolParserManager.get_tool_parser("lfm2")(lfm2_tokenizer)

496 full_text = (

497 f"Let me check. {TOOL_CALL_START}[{SIMPLE_FUNCTION_OUTPUT}]{TOOL_CALL_END}"

@@ -512,8 +508,7 @@ def test_streaming_leading_block_and_trailing_in_single_delta(

508 lfm2_tokenizer: TokenizerLike,

509 ):

510 """Leading text + complete tool block + trailing text in one delta.

515 - Both leading and trailing content must be preserved.

516 - """

511 + Both leading and trailing content must be preserved."""

512 tool_parser: ToolParser = ToolParserManager.get_tool_parser("lfm2")(lfm2_tokenizer)

513 full_text = (

514 "Let me check. "

@@ -534,8 +529,7 @@ def test_echoed_tool_call_body_not_leaked_to_content(

529 """LFM2 sometimes emits the tool call body again after the first

530 <|tool_call_end|>, capped with a second <|tool_call_end|>. The

531 echoed body must not surface as assistant content — neither in

537 - streaming nor non-streaming paths.

538 - """

532 + streaming nor non-streaming paths."""

533 tool_parser: ToolParser = ToolParserManager.get_tool_parser("lfm2")(lfm2_tokenizer)

534 body = (

535 "[grocery.orderIngredients("

@@ -570,8 +564,7 @@ def test_streaming_char_by_char_multi_dict_list(lfm2_tokenizer: TokenizerLike):

564 character at a time. Every prefix lands in some partial-parse state

565 (mid-key, mid-value, open quote inside dict, empty dict, etc.). The

566 parser must not raise — incomplete prefixes should silently wait for

573 - more text instead of logging exceptions.

574 - """

567 + more text instead of logging exceptions."""

568 tool_parser: ToolParser = ToolParserManager.get_tool_parser("lfm2")(lfm2_tokenizer)

569 full_text = (

570 f"{TOOL_CALL_START}[grocery.orderIngredients("

@@ -602,8 +595,7 @@ def test_streaming_char_by_char_multi_dict_list(lfm2_tokenizer: TokenizerLike):

595 def test_streaming_dotted_name_in_single_delta(lfm2_tokenizer: TokenizerLike):

596 """A pythonic call with a dotted/attribute function name (e.g.

597 ``domain.method(arg=...)``) must be parsed correctly in streaming mode

605 - just as in non-streaming mode.

606 - """

598 + just as in non-streaming mode."""

599 tool_parser: ToolParser = ToolParserManager.get_tool_parser("lfm2")(lfm2_tokenizer)

600 full_text = f"{TOOL_CALL_START}[{DOTTED_NAME_FUNCTION_OUTPUT}]{TOOL_CALL_END}"

601

@@ -619,8 +611,7 @@ def test_adjust_request_disables_skip_special_tokens(

611 """When tools are present, the parser must force

612 ``skip_special_tokens=False`` so the engine does not strip the

613 <|tool_call_start|>/<|tool_call_end|> sentinels before they reach the

622 - parser.

623 - """

614 + parser."""

615 from vllm.entrypoints.openai.chat_completion.protocol import (

616 ChatCompletionRequest,

617 )

tests/tool_use/test_muse_glimmer.py

+1 -2

@@ -220,8 +220,7 @@ def test_reasoning_then_parallel_calls():

220

221 def _stream(raw: str, chunk: int):

222 """Feed ``raw`` incrementally in ``chunk``-char steps through BOTH streaming

223 - parsers; return (reasoning, content, tool_calls).

224 - """

223 + parsers; return (reasoning, content, tool_calls)."""

224 reasoning, content, toolcalls = [], [], []

225 prev = ""

226 i = 0

tests/tool_use/test_tool_choice_required.py

+1 -2

@@ -376,8 +376,7 @@ WEB_SEARCH_TOOL = WebSearchTool(type="web_search")

376

377 class TestNonFunctionToolsSkipped:

378 """Non-function tools (web_search, etc.) must be silently skipped

379 - by the tool-schema utilities instead of raising TypeError.

380 - """

379 + by the tool-schema utilities instead of raising TypeError."""

380

381 def test_find_tool_properties_skips_web_search(self):

382 tools = [WEB_SEARCH_TOOL, FUNCTION_TOOL]

tests/transformers_utils/processors/test_glm5next.py

+2 -4

@@ -330,8 +330,7 @@ def test_sample_frames_fps_interval(video_processor):

330

331 def test_defaults_mirror_checkpoint_config():

332 """Bare instantiation matches the checkpoint's ``processor_config.json``

333 - token-budget style defaults; ``size`` carries no budget.

334 - """

333 + token-budget style defaults; ``size`` carries no budget."""

334 image_processor = Glm5NextImageProcessor()

335 assert image_processor.patch_expand_factor == 1

336 assert image_processor.min_image_tokens == 16

@@ -365,8 +364,7 @@ def test_missing_token_budgets_rejected():

364

365 def test_video_config_fields_land():

366 """fps_interval / max_frame_count_dynamic from the dedicated config

368 - shape sampling without any request overrides.

369 - """

367 + shape sampling without any request overrides."""

368 proc = Glm5NextVideoProcessor(

369 fps_interval=4,

370 max_frame_count_dynamic=32,

tests/transformers_utils/test_config.py

+2 -4

@@ -188,8 +188,7 @@ def test_model_config_generation_fallback_forwards_code_revision():

188

189 def test_safetensors_metadata_of_repo_without_safetensors():

190 """A repo storing its weights in another format is an answer, not a failure,

191 - so it must not be retried.

192 - """

191 + so it must not be retried."""

192 from huggingface_hub.errors import LocalEntryNotFoundError, NotASafetensorsRepoError

193

194 get_safetensors_metadata = MagicMock(

@@ -228,8 +227,7 @@ def test_mrope_num_dims(section_key, mrope_section, expected_num_dims):

227 @pytest.mark.parametrize("section_name", ["mrope_section", "xdrope_section"])

228 def test_mrope_num_dims_from_config_attribute(section_name):

229 """Some configs expose the section as an attribute rather than under

231 - `rope_parameters`.

232 - """

230 + `rope_parameters`."""

231 config = PretrainedConfig()

232 setattr(config, section_name, [16, 16, 16, 16])

233

tests/transformers_utils/test_hf_overrides_model_type.py

+1 -2

@@ -22,8 +22,7 @@ def test_hf_overrides_model_type_returns_correct_config_class():

22 """When hf_overrides sets model_type to a registered custom type whose

23 checkpoint has a *different* model_type on disk, get_config() must return

24 an instance of the registered config class — not the class that matches

25 - the on-disk model_type.

26 - """

25 + the on-disk model_type."""

26 # Register the custom config

27 _CONFIG_REGISTRY["test_custom_model"] = _TestCustomConfig

28

tests/utils.py

+2 -4

@@ -72,8 +72,7 @@ FP8_DTYPE = current_platform.fp8_dtype()

72 def prewarm_hf_cache(assets: list[tuple[str, str]]) -> None:

73 """Pre-populate the HF cache for (repo_id, filename) pairs that upstream

74 trust_remote_code modules would otherwise fetch from third-party CDNs

75 - (often unreachable from US-based CI).

76 - """

75 + (often unreachable from US-based CI)."""

76 if HF_HUB_OFFLINE:

77 return

78 for repo_id, filename in assets:

@@ -893,8 +892,7 @@ class RemoteOpenAIServerCustom(RemoteOpenAIServer):

892 max_wait_seconds: float | None = None,

893 ) -> None:

894 """Store custom child process function then invoke superclass

896 - constructor which will indirectly launch it.

897 - """

895 + constructor which will indirectly launch it."""

896 self.child_process_fxn = child_process_fxn

897 super().__init__(

898 model=model,

tests/utils_/test_gpu_sync_debug.py

+5 -10

@@ -59,8 +59,7 @@ def test_with_env_set(monkeypatch, mode):

59 @create_new_process_for_each_test()

60 def test_other_threads_are_not_policed(monkeypatch):

61 """A background thread that syncs deliberately must not be broken by the

62 - check being armed on the thread running the decorated function.

63 - """

62 + check being armed on the thread running the decorated function."""

63 monkeypatch.setattr(gsd, "_SYNC_CHECK_MODE", "error")

64 monkeypatch.setattr(gsd, "_sync_check_enabled", True)

65

@@ -85,8 +84,7 @@ def test_other_threads_are_not_policed(monkeypatch):

84 def test_allow_on_other_thread_does_not_disarm(monkeypatch):

85 """`gpu_sync_allowed()` on one thread must not suppress the check on

86 another. It is scoped by ContextVar rather than torch's process-global

88 - sync debug mode, which a previous implementation mutated.

89 - """

87 + sync debug mode, which a previous implementation mutated."""

88 monkeypatch.setattr(gsd, "_SYNC_CHECK_MODE", "error")

89 monkeypatch.setattr(gsd, "_sync_check_enabled", True)

90

@@ -115,8 +113,7 @@ def test_allow_on_other_thread_does_not_disarm(monkeypatch):

113 def test_suppressing_works_while_compiling(monkeypatch):

114 """`_suppressing` wraps torch compile entry points, which run with

115 `torch.compiler.is_compiling()` true. `gpu_sync_allowed()` deliberately

118 - no-ops in that state, so `_suppressing` must not route through it.

119 - """

116 + no-ops in that state, so `_suppressing` must not route through it."""

117 monkeypatch.setattr(gsd, "_SYNC_CHECK_MODE", "error")

118 monkeypatch.setattr(gsd, "_sync_check_enabled", True)

119 # Emulate being inside a torch compile, as inductor passes are.

@@ -148,8 +145,7 @@ def test_copy_checker_is_traceable_while_compiling(monkeypatch):

145 def test_sync_debug_mode_restored_after_checked_call(monkeypatch):

146 """The mode is armed only for the duration of a checked call. Leaving it

147 on process-wide made every sync outside a checked region emit a

151 - `UserWarning` whenever our handler was not the installed one.

152 - """

148 + `UserWarning` whenever our handler was not the installed one."""

149 monkeypatch.setattr(gsd, "_SYNC_CHECK_MODE", "error")

150 monkeypatch.setattr(gsd, "_sync_check_enabled", True)

151

@@ -294,8 +290,7 @@ def test_genuinely_async_transfers_pass(monkeypatch):

290 asynchronous in both directions -- even with a dtype conversion, which

291 runs GPU-side, and even permuted (e.g. transposed), which uses pitched

292 cudaMemcpy2D/3DAsync; D2H `Tensor.to` allocates a pinned destination.

297 - None of these may be flagged.

298 - """

293 + None of these may be flagged."""

294 monkeypatch.setattr(gsd, "_SYNC_CHECK_MODE", "error")

295 monkeypatch.setattr(gsd, "_sync_check_enabled", True)

296 gsd._install_copy_checkers()

tests/utils_/test_network_utils.py

+5 -10

@@ -29,8 +29,7 @@ def test_get_file_store_init_method_is_unique():

29

30 def _call_with_timeout(func, timeout: float = 10.0):

31 """Run func in a daemon thread so a livelock regression fails the test

32 - quickly instead of hanging the suite.

33 - """

32 + quickly instead of hanging the suite."""

33 result: dict = {}

34

35 def target():

@@ -97,8 +96,7 @@ def test_get_open_port_skips_reserved_dp_master_ports(

96 monkeypatch: pytest.MonkeyPatch,

97 ):

98 """VLLM_PORT == VLLM_DP_MASTER_PORT must not livelock by returning the

100 - same free-but-reserved port forever (upstream issue #50024).

101 - """

99 + same free-but-reserved port forever (upstream issue #50024)."""

100 base = 42123

101 with monkeypatch.context() as m:

102 m.setenv("VLLM_PORT", str(base))

@@ -111,8 +109,7 @@ def test_get_open_ports_list_skips_reserved_dp_master_ports(

109 monkeypatch: pytest.MonkeyPatch,

110 ):

111 """The VLLM_PORT band scan must not hand out ports inside the window

114 - reserved for the data parallel master process.

115 - """

112 + reserved for the data parallel master process."""

113 base = 42223

114 reserved = range(base, base + 10)

115 with monkeypatch.context() as m:

@@ -128,8 +125,7 @@ def test_get_open_port_ephemeral_skips_reserved_range(

125 monkeypatch: pytest.MonkeyPatch,

126 ):

127 """An ephemeral port that happens to fall in the reserved range is

131 - replaced by a rescan starting past the range.

132 - """

128 + replaced by a rescan starting past the range."""

129 base = 42323

130 with monkeypatch.context() as m:

131 m.delenv("VLLM_PORT", raising=False)

@@ -148,8 +144,7 @@ def test_get_open_port_ephemeral_without_dp_master_port(

144 monkeypatch: pytest.MonkeyPatch,

145 ):

146 """Without VLLM_PORT and VLLM_DP_MASTER_PORT, an ephemeral port is

151 - returned as before.

152 - """

147 + returned as before."""

148 with monkeypatch.context() as m:

149 m.delenv("VLLM_PORT", raising=False)

150 m.delenv("VLLM_DP_MASTER_PORT", raising=False)

tests/utils_/test_numa_utils.py

+6 -12

@@ -157,8 +157,7 @@ def test_pct_binding_filters_cpus(monkeypatch):

157 )

158 def test_pct_binding_fires_on_every_capable_sku(monkeypatch, sku, expected_cpus):

159 """Each SKU in ``_PCT_CAPABLE_SKUS`` engages the gate at its own

160 - expected ``highest_perf`` and uses its own priority-core stride.

161 - """

160 + expected ``highest_perf`` and uses its own priority-core stride."""

161 sku_config = numa_utils._PCT_CAPABLE_SKUS[sku]

162 _patch_pct_gates(

163 monkeypatch,

@@ -243,8 +242,7 @@ def test_get_numactl_args_engine_core_baseline_spans_shard_numa_nodes():

242 def test_get_numactl_args_engine_core_pct_spans_shard_numa_nodes(monkeypatch):

243 """PCT: EngineCore for a multi-NUMA shard binds to the union of priority

244 cores across all shard nodes, so worker `--physcpubind` is always a

246 - subset of EngineCore's `cpus_allowed`.

247 - """

245 + subset of EngineCore's `cpus_allowed`."""

246 _patch_pct_gates(

247 monkeypatch,

248 model_match=True,

@@ -294,8 +292,7 @@ def test_get_numactl_args_engine_core_pct_external_launcher_spans_local_nodes(

292 """external_launcher (or multi-node-within-DP, or Ray) hits the

293 fallback branch. EngineCore must still span every local NUMA node

294 so it can mp-spawn its local workers without ``--physcpubind``

297 - strict-validation failures.

298 - """

295 + strict-validation failures."""

296 _patch_pct_gates(

297 monkeypatch,

298 model_match=True,

@@ -319,8 +316,7 @@ def test_get_numactl_args_engine_core_pct_external_launcher_spans_local_nodes(

316

317 def test_get_numactl_args_engine_core_baseline_multi_node_within_dp_spans_locals():

318 """Multi-node-within-DP fallback: bind EngineCore to all local NUMA

322 - nodes that the visible ``numa_bind_nodes`` reference.

323 - """

319 + nodes that the visible ``numa_bind_nodes`` reference."""

320 vllm_config = _make_config(

321 numa_bind=True,

322 numa_bind_nodes=[0, 0, 0, 0, 1, 1, 1, 1],

@@ -343,8 +339,7 @@ def test_get_numactl_args_engine_core_skips_user_cpu_list(monkeypatch):

339 ``--physcpubind`` spawns need. We fall back to ``--cpunodebind`` over

340 the shard's NUMA nodes instead. PCT auto-detect is also bypassed when

341 the user is explicit (its priority-core union may not be a superset

346 - of the user's per-worker cores).

347 - """

342 + of the user's per-worker cores)."""

343 _patch_pct_gates(monkeypatch, model_match=True, highest_perf=46)

344 vllm_config = _make_config(

345 numa_bind=True,

@@ -397,8 +392,7 @@ def test_get_numactl_args_requires_detectable_nodes(monkeypatch):

392 def test_configure_subprocess_rejects_unknown_process_kind():

393 """configure_subprocess only knows 'worker' and 'EngineCore'; anything

394 else must raise ValueError instead of silently routing to the worker

400 - path.

401 - """

395 + path."""

396 vllm_config = _make_config(numa_bind=True, numa_bind_nodes=[0])

397 with (

398 pytest.raises(ValueError, match="process_kind"),

tests/v1/attention/test_attention_backends.py

+1 -2

@@ -1392,8 +1392,7 @@ def test_non_causal_backend_correctness(

1392 model: str,

1393 ):

1394 """Test backend's correctness with non-causal (bidirectional) decoder

1395 - attention, as used by DFlash speculative decoding.

1396 - """

1395 + attention, as used by DFlash speculative decoding."""

1396

1397 def bidirectional_mask_mod(

1398 b: torch.Tensor,

tests/v1/attention/test_attention_backends_selection.py

+1 -2

@@ -115,8 +115,7 @@ def test_mamba_layers_have_unified_interface(

115 layer_class, expected_backend, expected_mamba_type

116 ):

117 """Test that all Mamba layers have the unified get_attn_backend

118 - interface.

119 - """

118 + interface."""

119 assert hasattr(layer_class, "get_attn_backend"), (

120 f"{layer_class.__name__} should have get_attn_backend method"

121 )

tests/v1/attention/test_attention_splitting.py

+1 -2

@@ -164,8 +164,7 @@ def apply_split_decodes_and_prefills(

164 treat_short_extends_as_decodes: bool = True,

165 ):

166 """Helper function to apply split_decodes_and_prefills and return

167 - the results.

168 - """

167 + the results."""

168 device = torch.device("cpu")

169 seq_lens = [10 * (i + 1) for i in range(len(query_lens))]

170 common_metadata = create_common_attn_metadata(

tests/v1/attention/test_dcp_a2a_pack_mask.py

+1 -2

@@ -1,8 +1,7 @@

1 # SPDX-License-Identifier: Apache-2.0

2 # SPDX-FileCopyrightText: Copyright contributors to the vLLM project

3 """The empty-shard mask fused into the DCP A2A pack kernel must be bit-exact

4 -with the eager ``mask_dcp_empty_shards_`` pass it replaces.

5 -"""

4 +with the eager ``mask_dcp_empty_shards_`` pass it replaces."""

5

6 import pytest

7 import torch

tests/v1/attention/test_dspark_noncausal_sparse_mla.py

+1 -2

@@ -495,8 +495,7 @@ def test_dspark_noncausal_sparse_mla_matches_sdpa(

495 num_heads,

496 ):

497 """Non-causal (window ∪ block, future-pointing) per-token indices must match

498 - a dense SDPA reference over the same indices, for both sparse-MLA backends.

499 - """

498 + a dense SDPA reference over the same indices, for both sparse-MLA backends."""

499 _skip_if_backend_unavailable(backend_cls, kv_cache_dtype, block_size)

500

501 window, block, topk_width = _DSPARK_CONFIGS[config_name]

tests/v1/attention/test_flashinfer_mla_sparse_sm90.py

+1 -2

@@ -231,8 +231,7 @@ def test_plan_uses_state_params(monkeypatch):

231

232 def test_kv_lens_host_formula():

233 """Per-row host lengths: context == position + 1; capped at

234 - index_topk + trailing-pool remainder past the sparse threshold.

235 - """

234 + index_topk + trailing-pool remainder past the sparse threshold."""

235 builder = object.__new__(FlashInferMLASparseSM90Builder)

236 builder._index_topk = 2048

237 builder._index_kpool = 4

tests/v1/attention/test_gdn_metadata_builder.py

+1 -2

@@ -187,8 +187,7 @@ def test_gdn_build_classification(test_case: GDNBuildTestCase):

187 def test_has_initial_state_after_reclassification():

188 """After reclassification, num_prefills > 0 so the prefill kernel path

189 should compute has_initial_state. For the reclassified request with

190 - context_lens > 0, the corresponding entry must be True.

191 - """

190 + context_lens > 0, the corresponding entry must be True."""

191 builder = _create_gdn_builder(num_speculative_tokens=2)

192 batch = BatchSpec(seq_lens=[65, 20], query_lens=[1, 3])

193 meta = _build(builder, batch, num_decode_draft_tokens=[-1, 2])

tests/v1/attention/test_indexer_dcp_localize.py

+6 -12

@@ -56,8 +56,7 @@ def _ref_stable_topk_from_candidates_fp64(

56 ) -> torch.Tensor:

57 """Pure-PyTorch reference for the CuteDSL stable-topk selector order

58 (score desc, then lowest global token id). Selects the same SET as the

59 - kernel; only the set is compared in tests.

60 - """

59 + kernel; only the set is compared in tests."""

60 num_rows, num_candidates = candidate_scores.shape

61 device = candidate_scores.device

62 select_k = min(k, num_candidates)

@@ -122,8 +121,7 @@ def _dcp_lse_merge(

121 class _FakeDCPGroup:

122 """Single-process stand-in: ``all_gather`` returns the pre-built

123 concatenation of every rank's packed ``(score, global_id)`` candidates,

125 - mirroring the one packed all-gather the merge issues.

126 - """

124 + mirroring the one packed all-gather the merge issues."""

125

126 def __init__(self, gathered_packed: torch.Tensor) -> None:

127 self.gathered_packed = gathered_packed

@@ -209,8 +207,7 @@ def _merge_local_topks_global_with_fake_dcp(

207 """Run ``_merge_dcp_topk_global`` for every rank against a faked

208 all-gather and return each rank's global-index result (all ranks should

209 agree). The fake pre-builds the packed candidate concatenation the merge's

212 - single ``all_gather`` would return.

213 - """

210 + single ``all_gather`` would return."""

211 # The merge is now CuteDSL-only (no PyTorch fallback), so it runs the real

212 # Triton pack + CuteDSL selector kernels even behind the faked all-gather.

213 if not current_platform.is_cuda() or not has_cutedsl():

@@ -712,8 +709,7 @@ def test_dcp_filter_compaction_matches_reference(

709 """In-kernel compaction must, for every row, produce exactly the rank-owned

710 physical slots packed into [0, valid_count) with -1 in the tail -- the same

711 SET a reference filter + sort/gather produces. Uses wide rows (> BLOCK_N

715 - valid slots), with interior -1 gaps.

716 - """

712 + valid slots), with interior -1 gaps."""

713 device = torch.device("cuda")

714 torch.manual_seed(7)

715 dcp_size = 2

@@ -871,8 +867,7 @@ def test_decode_topk_pads_surplus_with_negative_one():

867 so a non-(-1) pad would be silently attended. This is the common case under

868 DCP, where each rank's local seq_len = global / world is usually << topk.

869 Checked on the *set* of valid indices (the kernel may order them by score,

874 - not position).

875 - """

870 + not position)."""

871 torch.manual_seed(0)

872 device = torch.device("cuda")

873 topk = 8

@@ -918,8 +913,7 @@ def test_sparse_decode_dcp_short_context_matches_non_dcp():

913 """End-to-end DCP decode where the global seq_len < topk (so every rank's

914 local top-k is surplus-padded). Exercises the kernel surplus -> merge mask

915 -> global top-k -> physical localize -> LSE merge chain for the common

921 - short-context decode case, vs the non-DCP reference.

922 - """

916 + short-context decode case, vs the non-DCP reference."""

917 torch.manual_seed(4)

918 device = torch.device("cuda")

919 world = 2

tests/v1/attention/test_kpool_tail_slot_mapping.py

+7 -14

@@ -58,8 +58,7 @@ def test_tail_backend_layout_matches_kernel_pointer_arithmetic():

58 def make_tail_block_table(own_blocks, width=64):

59 """Tail-group block table as BlockTables produces it: column 0 holds the

60 request's single KpoolTailManager block, the remaining columns are never

61 - written and stay zero.

62 - """

61 + written and stay zero."""

62 bt = torch.zeros(len(own_blocks), width, dtype=torch.int32)

63 bt[:, 0] = torch.tensor(own_blocks, dtype=torch.int32)

64 return bt

@@ -67,8 +66,7 @@ def make_tail_block_table(own_blocks, width=64):

66

67 def legacy_generic_tail_slots(block_table, query_start_loc, positions):

68 """Reference of the generic ``_compute_slot_mappings_kernel`` arithmetic

70 - (block_table.py:305-313) applied to the tail group's table.

71 - """

69 + (block_table.py:305-313) applied to the tail group's table."""

70 slots = []

71 for req in range(block_table.shape[0]):

72 for i in range(query_start_loc[req], query_start_loc[req + 1]):

@@ -109,8 +107,7 @@ def make_batch(per_req_positions, padded_len=None):

107

108 def test_legacy_generic_mapping_collapses_onto_block_zero():

109 """The bug: with the manager's 1-column block table, the generic kernel

112 - maps every pos >= kpool onto tail block 0, and distinct requests collide.

113 - """

110 + maps every pos >= kpool onto tail block 0, and distinct requests collide."""

111 own_blocks = [5, 9]

112 per_req = [list(range(10)), list(range(12))] # prompts of len 10 and 12

113 positions, qsl, _, num_actual, num_reqs = make_batch(per_req)

@@ -140,8 +137,7 @@ def test_legacy_generic_mapping_collapses_onto_block_zero():

137 def test_circular_mapping_isolates_requests():

138 """The fix: every token lands in its own request's block at pos % kpool,

139 and no slot is ever shared by two different requests (slots do recur

143 - within a request every kpool positions -- that is the circular design).

144 - """

140 + within a request every kpool positions -- that is the circular design)."""

141 own_blocks = [5, 9]

142 per_req = [list(range(10)), list(range(12))]

143 positions, qsl, slot_mapping, num_actual, num_reqs = make_batch(per_req)

@@ -169,8 +165,7 @@ def test_circular_mapping_isolates_requests():

165 def test_circular_mapping_matches_generic_for_short_requests(prompt_len):

166 """For pos < kpool the generic kernel already picks the own block, so the

167 two mappings agree while every position fits the request's first block

172 - (single-request behavior is unchanged).

173 - """

168 + (single-request behavior is unchanged)."""

169 own_blocks = [7]

170 per_req = [list(range(prompt_len))]

171 positions, qsl, slot_mapping, num_actual, num_reqs = make_batch(per_req)

@@ -280,8 +275,7 @@ class TailRingMirror:

275 """Mirror of _kpool_tail_seed_kernel / _kpool_decode_update_batched_kernel

276 addressing: block = tail_slot // kpool, ring offset = pos % kpool; a pool

277 completing at pos reads ring slots (pool_start + s) % kpool and uses the

283 - current token's own K/score for the last member.

284 - """

278 + current token's own K/score for the last member."""

279

280 def __init__(self, num_blocks, kpool=KPOOL):

281 self.kpool = kpool

@@ -324,8 +318,7 @@ def tail_slot_for(mapping, req, pos, own_block):

318

319 def run_scenario(mapping, interleave):

320 """Requests A (block 5, prompt len 9) and B (block 9, prompt len 11)

327 - decode concurrently; returns A's boundary pool [8, 9, 10, 11].

328 - """

321 + decode concurrently; returns A's boundary pool [8, 9, 10, 11]."""

322 ring = TailRingMirror(num_blocks=16)

323 blocks = {"A": 5, "B": 9}

324 prompts = {"A": 9, "B": 11}

tests/v1/attention/test_mamba_update_block_table.py

+7 -14

@@ -28,8 +28,7 @@ def _make_vllm_config(

28 block_size: int | None = None,

29 ):

30 """Create a minimal mock VllmConfig with only the fields the builder

31 - accesses, avoiding any model download / HF config inspection.

32 - """

31 + accesses, avoiding any model download / HF config inspection."""

32 speculative_config = (

33 SimpleNamespace(

34 num_speculative_tokens=num_speculative_tokens,

@@ -75,8 +74,7 @@ def test_mamba_single_token_prompt_runs_as_prefill():

74 def test_update_block_table_copies_block_idx_to_persistent_buffers():

75 """update_block_table() must write block_idx tensors to the current

76 builder's persistent buffers, not leave them pointing to a different

78 - builder's buffers.

79 - """

77 + builder's buffers."""

78 block_size = 16

79 max_model_len = 256

80 num_reqs = 4

@@ -180,8 +178,7 @@ def test_state_indices_tensor_d_includes_num_speculative_blocks():

178 bug 1: with mamba_cache_mode='all' and speculative decoding enabled,

179 the cudagraph buffer for state_indices_tensor_d must allocate the same

180 per-request column count as the runtime block table, which includes

183 - num_speculative_blocks trailing scratch columns.

184 - """

181 + num_speculative_blocks trailing scratch columns."""

182 block_size = 16

183 max_model_len = 256

184 max_num_seqs = 4

@@ -214,8 +211,7 @@ def test_block_idx_cudagraph_capture_padded_by_num_reqs():

211 bug 2: with mamba_cache_mode='all' and spec decode, _update_metadata_for

212 _cudagraph_capture must slice block_idx_last_{scheduled,computed}_token

213 by the request count (padded_bs == num_reqs), not by num_decode_tokens.

217 - Past num_decodes, the slice must be zero-filled.

218 - """

214 + Past num_decodes, the slice must be zero-filled."""

215 block_size = 16

216 max_model_len = 256

217 max_num_seqs = 8

@@ -295,8 +291,7 @@ def test_block_idx_prev_step_persistent_buffer_allocated():

291 block_idx_last_scheduled_token_prev_step as a persistent buffer with the

292 same shape as the existing block_idx_last_{scheduled,computed}_token

293 buffers, so cudagraph capture records a stable pointer for the prev-step

298 - input anchor consumed by mamba_mixer2's input gather.

299 - """

294 + input anchor consumed by mamba_mixer2's input gather."""

295 block_size = 16

296 max_model_len = 256

297 max_num_seqs = 8

@@ -324,8 +319,7 @@ def test_block_idx_prev_step_persistent_buffer_allocated():

319

320 def test_block_idx_prev_step_persistent_buffer_skipped_without_spec_decode():

321 """Without spec decode, the prev-step buffer is unused and must not be

327 - allocated — the input anchor reduces to last_computed_token.

328 - """

322 + allocated — the input anchor reduces to last_computed_token."""

323 block_size = 16

324 max_model_len = 256

325 max_num_seqs = 8

@@ -349,8 +343,7 @@ def test_block_idx_prev_step_cudagraph_capture_uses_persistent_buffer():

343 """_update_metadata_for_cudagraph_capture must copy the prev-step anchor

344 into the builder's persistent buffer (so cudagraph replay reads from the

345 same underlying memory), pad past num_decodes with zero, and return a

352 - slice of the persistent buffer in the metadata.

353 - """

346 + slice of the persistent buffer in the metadata."""

347 block_size = 16

348 max_model_len = 256

349 max_num_seqs = 8

tests/v1/attention/test_mla_backends.py

+1 -2

@@ -171,8 +171,7 @@ def test_glm5_flashinfer_masked_mha_routing(

171 def test_concat_k_nope_k_pe_matches_torch_cat(qk_rope_head_dim):

172 """The K concat used by the MLA prefill context loop must equal torch.cat of

173 k_nope with the broadcast k_pe; with no RoPE part it returns k_nope itself

174 - instead of allocating and copying.

175 - """

174 + instead of allocating and copying."""

175 torch.manual_seed(0)

176 num_tokens, num_heads, qk_nope_head_dim = 5, 4, 256

177 k_nope = torch.randn(num_tokens, num_heads, qk_nope_head_dim, dtype=torch.bfloat16)

tests/v1/attention/test_mla_prefill_quant_output.py

+3 -6

@@ -134,8 +134,7 @@ def test_flash_attn_registers_warmup_only_for_fa4(

134

135 def test_flash_attn_prefill_backend_signature_accepts_fused_kwargs():

136 """run_prefill_new_tokens must accept out/output_scale so the direct

137 - (non-**kwargs) call in forward_mha type- and runtime-checks.

138 - """

137 + (non-**kwargs) call in forward_mha type- and runtime-checks."""

138 import inspect

139

140 params = inspect.signature(

@@ -151,8 +150,7 @@ def test_flash_attn_prefill_backend_signature_accepts_fused_kwargs():

150

151 def test_mla_impl_forward_mha_accepts_output_scale():

152 """The abstract MLA impl forward_mha must carry output_scale so every

154 - override (and the unconditional forward_impl call) stays compatible.

155 - """

153 + override (and the unconditional forward_impl call) stays compatible."""

154 import inspect

155

156 from vllm.v1.attention.backend import MLAAttentionImpl

@@ -179,8 +177,7 @@ _FUSED_FP8_SKIP = _fused_fp8_skip_reason()

177 def test_fa4_fused_fp8_output_matches_post_quant(default_vllm_config):

178 """FA4's fused FP8 write (output_scale, flash-attention#135) must match the

179 bf16-attention + standalone static-FP8-quant path it replaces, since

182 - production uses the same output_scale for both.

183 - """

180 + production uses the same output_scale for both."""

181 from vllm.model_executor.layers.quantization.input_quant_fp8 import QuantFP8

182 from vllm.model_executor.layers.quantization.utils.quant_utils import GroupShape

183 from vllm.platforms import current_platform

tests/v1/attention/test_mla_prefill_selector.py

+1 -2

@@ -315,8 +315,7 @@ class TestBackendValidation:

315

316 def test_flash_attn_accepts_glm53_flash_nope_dimensions(self):

317 """(256, 0, 256) runs the same FA kernels as GLM-5's (192, 64, 256); a

318 - RoPE-carrying 320-wide query does not.

319 - """

318 + RoPE-carrying 320-wide query does not."""

319 try:

320 from vllm.v1.attention.backends.mla.prefill.flash_attn import (

321 FlashAttnPrefillBackend,

tests/v1/attention/test_replayssm_metadata_builder.py

+1 -2

@@ -257,8 +257,7 @@ def test_replayssm_write_pos(case: ReplaySSMBuildCase):

257

258 def test_resumed_request_differs_from_fresh():

259 """Same token count, different decode_base: fresh (base 100) -> write_pos 5,

260 - resumed (base 105) -> write_pos 0.

261 - """

260 + resumed (base 105) -> write_pos 0."""

261 builder = _create_replayssm_builder(16)

262 batch = BatchSpec(seq_lens=[106, 106], query_lens=[1, 1])

263 common = create_common_attn_metadata(batch, BLOCK_SIZE, DEVICE).replace(

tests/v1/attention/test_rocm_attention_backends_selection.py

+1 -2

@@ -517,8 +517,7 @@ def test_auto_selection_for_kv_connector(

517 aiter_found, expected_backend, mock_vllm_config, mock_get_cdna_version

518 ):

519 """Auto-selection with a KV connector and AITER enabled resolves to unified attn,

520 - and to triton attn if AITER not enabled.

521 - """

520 + and to triton attn if AITER not enabled."""

521 from vllm.platforms.rocm import RocmPlatform

522

523 with patch(

tests/v1/attention/test_sparse_indexer_decode_seq_lens.py

+4 -8

@@ -102,8 +102,7 @@ def expected_row_seq_lens(per_req_positions, batch_size, next_n):

102

103 def test_legacy_flat_layout_misaligns_and_bleeds():

104 """The bug: ``positions[:n] + 1`` with padded-row coordinates reads other

105 - requests' positions and prefill positions.

106 - """

105 + requests' positions and prefill positions."""

106 decode_lens, per_req, positions = make_non_uniform_batch()

107 batch_size = decode_lens.shape[0]

108 next_n = int(decode_lens.max())

@@ -138,8 +137,7 @@ def test_helper_uniform_layout_matches_flat_slice():

137

138 def test_helper_padded_layout_per_row():

139 """Non-uniform batches map every padded row to its own token's position;

141 - pad rows get 0 (empty tail).

142 - """

140 + pad rows get 0 (empty tail)."""

141 decode_lens, per_req, positions = make_non_uniform_batch()

142 out = _decode_topk_seq_lens(positions, decode_lens, 8, 3, 4, requires_padding=True)

143 expected = expected_row_seq_lens(per_req, 3, 4)

@@ -151,8 +149,7 @@ def test_helper_padded_layout_per_row():

149 def expand_tail_region(dec_seq):

150 """Run the production pure-torch expand + append pair (the fused kernel

151 replicates it exactly on the identity path) and return the tail columns

154 - [TOPK_TOKENS, TOPK_TOKENS + KPOOL - 1).

155 - """

152 + [TOPK_TOKENS, TOPK_TOKENS + KPOOL - 1)."""

153 rows = dec_seq.shape[0]

154 pool_ids = torch.arange(SELECT_K, dtype=torch.int64).expand(rows, SELECT_K)

155 valid = torch.ones_like(pool_ids, dtype=torch.bool)

@@ -166,8 +163,7 @@ def expand_tail_region(dec_seq):

163 def test_tail_expansion_legacy_vs_fixed():

164 """End-to-end tail consequence: the fixed mapping appends exactly the

165 request's trailing incomplete pool; the legacy flat slice drops real tail

169 - tokens and emits indices far past the request's own sequence.

170 - """

166 + tokens and emits indices far past the request's own sequence."""

167 decode_lens, per_req, positions = make_non_uniform_batch()

168 n = 12

169 fixed = _decode_topk_seq_lens(

tests/v1/attention/test_sparse_mla_backends.py

+9 -18

@@ -909,8 +909,7 @@ def test_index_group_convert_during_piecewise_capture():

909 runs in the following eager break. The convert's side stream must not

910 wait on ``logical_topk_ready`` there: an event recorded inside a captured

911 segment is graph-local, and an eager wait on it raises

912 - cudaErrorInvalidValue.

913 - """

912 + cudaErrorInvalidValue."""

913 device = torch.device(DEVICE_TYPE)

914 num_tokens, num_topk, num_requests, blocks_per_req, block_size = 8, 128, 4, 4, 16

915

@@ -1049,8 +1048,7 @@ def test_triton_convert_rejects_req_id_longer_than_token_indices():

1048 req_id but the output is allocated like token_indices, so a full-batch

1049 req_id combined with an MQA-subset token_indices wrote past the end of

1050 the output buffer. The wrapper must reject the length mismatch instead

1052 - of corrupting memory.

1053 - """

1051 + of corrupting memory."""

1052 device = torch.device(DEVICE_TYPE)

1053 num_topk_tokens = 128

1054 block_size = 64

@@ -1105,8 +1103,7 @@ def test_flashmla_forward_bf16_kv_slices_req_id_to_mqa_tokens():

1103 is active, forward_mqa only receives the leading decode tokens, but

1104 _forward_bf16_kv passed the full-batch req_id_per_token to the index

1105 conversion, making it write past the end of its output buffer. The call

1108 - site must slice req_id_per_token to the MQA tokens.

1109 - """

1106 + site must slice req_id_per_token to the MQA tokens."""

1107 device = torch.device(DEVICE_TYPE)

1108 num_topk_tokens = 128

1109 block_size = 64

@@ -1204,8 +1201,7 @@ def test_split_prefill_chunks(seq_lens, max_buf, expected):

1201 )

1202 def test_masked_mha_workspace_fits_single_request_boundary(max_query_len, expected):

1203 """A 32K prefill needs the default workspace exactly; shrinking it would

1207 - push a supported request onto MQA.

1208 - """

1204 + push a supported request onto MQA."""

1205 assert (

1206 _masked_mha_workspace_fits(

1207 batch_size=1,

@@ -1263,8 +1259,7 @@ def test_is_masked_mha_available_model_dims(

1259 ):

1260 """The allow-list gates masked MHA per exact model geometry: the DeepSeek-V3.2,

1261 GLM-5 and NoPE GLM-5.3-Flash layouts on an SM100-family GPU with FA4 and an

1266 - unquantized KV cache, nothing else.

1267 - """

1262 + unquantized KV cache, nothing else."""

1263 import vllm.model_executor.layers.attention.sparse_mla_attention as mod

1264

1265 monkeypatch.setattr(

@@ -3604,8 +3599,7 @@ def test_fp8_mixed_batch_dcp_neutralizes_empty_rows(monkeypatch):

3599 """A decode row whose top-k shard holds no local candidates (all -1) has

3600 undefined kernel out/lse; it must come back as (0, -inf), the identity of

3601 the cross-rank LSE merge, or a NaN would survive the merge even at zero

3607 - weight (0 * NaN = NaN).

3608 - """

3602 + weight (0 * NaN = NaN)."""

3603 num_tokens, num_heads, head_dim = 3, 2, 3

3604 q = torch.empty(num_tokens, num_heads, head_dim, device=DEVICE_TYPE)

3605 local_indices = torch.tensor(

@@ -3778,8 +3772,7 @@ def test_hisparse_fp8_prefill_gather_uses_dedicated_stream(monkeypatch):

3772 def test_sparse_impl_observes_repointed_indexer_buffer():

3773 """The MTP proposer repoints the draft's indexer at the target model's buffer

3774 after the backend impl is built, so the impl must resolve the buffer per read.

3781 - Snapshotting it in __init__ leaves the layer reading indices nothing writes.

3782 - """

3775 + Snapshotting it in __init__ leaves the layer reading indices nothing writes."""

3776 impl = object.__new__(FlashInferMLASparseImpl)

3777 own = torch.zeros(4, 8, dtype=torch.int32)

3778 target = torch.ones(4, 8, dtype=torch.int32)

@@ -3795,8 +3788,7 @@ def test_sparse_impl_observes_repointed_indexer_buffer():

3788

3789 def test_explicit_topk_buffer_supersedes_indexer():

3790 """Backbone skip-topk layers have no indexer, and the proposer also assigns the

3798 - shared buffer directly onto draft submodules.

3799 - """

3791 + shared buffer directly onto draft submodules."""

3792 impl = object.__new__(FlashInferMLASparseImpl)

3793 indexer = SimpleNamespace(

3794 topk_indices_buffer=torch.zeros(2, 2, dtype=torch.int32),

@@ -3811,8 +3803,7 @@ def test_explicit_topk_buffer_supersedes_indexer():

3803

3804 def test_sparse_mla_common_impl_resolves_buffer_lazily():

3805 """Guards the whole SparseMLACommonImpl family at once: a backend that

3814 - snapshots the buffer instead silently loses MTP buffer sharing.

3815 - """

3806 + snapshots the buffer instead silently loses MTP buffer sharing."""

3807 assert issubclass(SparseMLACommonImpl, SharedTopkIndicesBuffer)

3808 assert isinstance(SparseMLACommonImpl.topk_indices_buffer, property), (

3809 "topk_indices_buffer must stay a lazily-resolved property"

tests/v1/attention/test_trtllm_attention_integration.py

+3 -6

@@ -198,8 +198,7 @@ def _create_nvfp4_hnd_kv_cache(

198

199 def _run_trtllm_integration(batch_spec, kv_cache_dtype="auto", model_name=MODEL):

200 """Run TRTLLM attention through the full FlashInfer pipeline

201 - and compare against an SDPA reference.

202 - """

201 + and compare against an SDPA reference."""

202 set_random_seed(42)

203 device = torch.device(f"{DEVICE_TYPE}:0")

204

@@ -427,8 +426,7 @@ def _run_trtllm_integration(batch_spec, kv_cache_dtype="auto", model_name=MODEL)

426 def test_trtllm_gen_full_attention_integration(batch_spec_name: str):

427 """Test TRTLLM gen-full attention through the full FlashInfer

428 MetadataBuilder.build() -> FlashInferImpl.forward() pipeline,

430 - with real TRTLLM kernels on Blackwell.

431 - """

429 + with real TRTLLM kernels on Blackwell."""

430 _run_trtllm_integration(BATCH_SPECS[batch_spec_name])

431

432

@@ -439,8 +437,7 @@ def test_trtllm_gen_full_attention_integration(batch_spec_name: str):

437 @torch.inference_mode()

438 def test_trtllm_gen_nvfp4_kv_integration(batch_spec_name: str):

439 """Test TRTLLM attention with nvfp4 KV cache through the full

442 - FlashInfer MetadataBuilder.build() -> FlashInferImpl.forward() pipeline.

443 - """

440 + FlashInfer MetadataBuilder.build() -> FlashInferImpl.forward() pipeline."""

441 _run_trtllm_integration(

442 BATCH_SPECS[batch_spec_name],

443 kv_cache_dtype="nvfp4",

tests/v1/core/prefix_cache/test_partial_prefix_cache_hits.py

+17 -34

@@ -2,8 +2,7 @@

2 # SPDX-FileCopyrightText: Copyright contributors to the vLLM project

3 """Fine-grained partial prefix-cache hits for hybrid (full attention + mamba

4 "align") models: scheduler chunk splitting, partial tail registration, CoW

5 -on partial hits, and same-step deferral.

6 -"""

5 +on partial hits, and same-step deferral."""

6

7 from math import lcm

8 from types import SimpleNamespace

@@ -175,8 +174,7 @@ def test_mamba_align_split_partial_tail_schedule(dcp_world_size: int):

174 """Chunk ends with partial hits on: block-aligned chunks, one extra stop

175 at the prompt's last hash boundary (registering the partial tail), then

176 the remaining tokens. block=512, hash=32, prompt=10000, budget=8192:

178 - 0 -> 8192 -> 9728 -> 9984 -> 10000.

179 - """

177 + 0 -> 8192 -> 9728 -> 9984 -> 10000."""

178 block_size = 512

179 scheduler_block_size = block_size * dcp_world_size

180 hash_block_size = 32

@@ -744,8 +742,7 @@ def test_hash_aligned_query_end_uses_regular_partial_tail():

742

743 def test_external_mamba_hit_same_block_uses_running_cow_on_continue():

744 """An external mid-block hit must become a running request even when its

747 - first continuation does not need another Mamba block.

748 - """

745 + first continuation does not need another Mamba block."""

746 hash_block_size = 2

747 mamba_block_size = 4 * hash_block_size

748 kv_cache_config = KVCacheConfig(

@@ -817,8 +814,7 @@ def test_external_mamba_hit_same_block_uses_running_cow_on_continue():

814

815 def test_boundary_state_offloads_returns_cow_target():

816 """Boundary hand-offs expose aligned snapshots and the partial-tail CoW

820 - target, never the overwritten CoW source.

821 - """

817 + target, never the overwritten CoW source."""

818 hash_block_size = 2

819 block_size = 2 * hash_block_size

820 kv_cache_config = KVCacheConfig(

@@ -1004,8 +1000,7 @@ def test_connector_finish_registers_partial_tail_before_cleanup():

1000 def test_boundary_state_offload_dropped_when_request_freed_before_drain():

1001 """A hand-off recorded in the same scheduling pass as the request's death

1002 must not be drained: its release hook has already run, so draining would

1007 - leak a pinned block.

1008 - """

1003 + leak a pinned block."""

1004 hash_block_size = 2

1005 block_size = 2 * hash_block_size

1006 kv_cache_config = KVCacheConfig(

@@ -1052,8 +1047,7 @@ def test_boundary_state_offload_dropped_when_request_freed_before_drain():

1047

1048 def test_boundary_state_offloads_block_aligned_prompt():

1049 """A prompt ending on a block boundary registers no CoW partial tail; its

1055 - boundary state block is handed off as a snapshot instead (once).

1056 - """

1050 + boundary state block is handed off as a snapshot instead (once)."""

1051 hash_block_size = 2

1052 block_size = 2 * hash_block_size

1053 kv_cache_config = KVCacheConfig(

@@ -1107,8 +1101,7 @@ def test_boundary_state_offloads_block_aligned_prompt():

1101 def test_truncate_computed_blocks_preserves_sparse_prefix_positions():

1102 """truncate_computed_blocks slices each group by its own block size,

1103 keeps null placeholders in the retained prefix, and leaves the original

1110 - lookup result untouched (pure view, no refcount changes).

1111 - """

1104 + lookup result untouched (pure view, no refcount changes)."""

1105 hash_block_size = 2

1106 kv_cache_config = KVCacheConfig(

1107 num_blocks=24,

@@ -1163,8 +1156,7 @@ def test_truncate_computed_blocks_preserves_sparse_prefix_positions():

1156

1157 def test_truncate_computed_blocks_allows_short_mamba_group_only():

1158 """External state may replace a short Mamba hit, but other groups must

1166 - cover the aligned local endpoint.

1167 - """

1159 + cover the aligned local endpoint."""

1160 hash_block_size = 2

1161 kv_cache_config = KVCacheConfig(

1162 num_blocks=24,

@@ -1306,8 +1298,7 @@ def test_hybrid_mamba_partial_tail_owner_continue_preserves_later_hit():

1298 def test_hybrid_mamba_moved_partial_entry_defers_same_step_hit():

1299 """The owner's move re-arms the same-step guard: the moved entry is

1300 filled by this step's copy, and chained same-step copies read stale

1309 - sources, so a request hitting it in the move step must be deferred.

1310 - """

1301 + sources, so a request hitting it in the move step must be deferred."""

1302 hash_block_size = 2

1303 block_size = 2 * hash_block_size

1304 kv_cache_config = KVCacheConfig(

@@ -1597,8 +1588,7 @@ def test_hybrid_partial_hash_truncates_full_attention_hit_length():

1588

1589 def test_cow_retained_blocks_returned_for_release():

1590 """new_step_starts returns the CoW copy retentions instead of freeing

1600 - them; the scheduler owns releasing them once the copy has run.

1601 - """

1591 + them; the scheduler owns releasing them once the copy has run."""

1592 hash_block_size = 2

1593 block_size = 2 * hash_block_size

1594 kv_cache_config = KVCacheConfig(

@@ -1651,8 +1641,7 @@ def test_cow_retained_blocks_returned_for_release():

1641

1642 def test_free_cow_retained_blocks_defers_until_copy_step_processed():

1643 """Scheduler releases CoW retentions immediately when the copy's step has

1654 - been processed (or deferral is off), and defers them otherwise.

1655 - """

1644 + been processed (or deferral is off), and defers them otherwise."""

1645 from collections import deque

1646

1647 freed: list = []

@@ -1689,8 +1678,7 @@ def test_free_cow_retained_blocks_defers_until_copy_step_processed():

1678 def test_full_attention_eagle_drops_one_hash_unit():

1679 """With fine-grained partial hits, eagle rewinds the hit by one hash unit

1680 instead of a whole cache block: the tail block's KV is append-only, so it

1692 - still covers the reduced length and stays in the hit as a partial block.

1693 - """

1681 + still covers the reduced length and stays in the hit as a partial block."""

1682 from vllm.v1.core.block_pool import BlockPool

1683 from vllm.v1.core.single_type_kv_cache_manager import FullAttentionManager

1684

@@ -1766,8 +1754,7 @@ def test_hybrid_partial_hit_with_eagle_stays_within_group_blocks():

1754 """Regression: with eagle, the mamba group must not receive the eagle

1755 lookup margin — its finder never applies the drop, so it could return a

1756 hit past the blocks the (dropped) full-attention group covers, crashing

1769 - the consumer's CoW with block_idx >= len(req_blocks).

1770 - """

1757 + the consumer's CoW with block_idx >= len(req_blocks)."""

1758 hash_block_size = 2

1759 block_size = 2 * hash_block_size

1760 kv_cache_config = KVCacheConfig(

@@ -2369,8 +2356,7 @@ def test_retention_snapshots_handed_off_with_exact_block_ids():

2356 """Under sparse retention, each retained mamba boundary state block is

2357 handed to the connector with its exact block id. Connectors must not resolve

2358 align-mode state blocks positionally because the block table is not

2372 - append-only.

2373 - """

2359 + append-only."""

2360 hash_block_size = 2

2361 block_size = 4

2362 manager = make_kv_cache_manager(

@@ -2411,8 +2397,7 @@ def test_snapshot_handoff_dense_default_retention():

2397 """With dense (default) retention, every materialized mamba boundary

2398 state block is handed off — regular mamba-align + prefix-match-unit

2399 deployments get store-able boundary snapshots without setting a

2414 - prefix-cache retention interval.

2415 - """

2400 + prefix-cache retention interval."""

2401 hash_block_size = 2

2402 block_size = 4

2403 manager = make_kv_cache_manager(

@@ -2444,8 +2429,7 @@ def test_snapshot_handoff_dense_default_retention():

2429

2430 def _run_chunked_prefill(manager, req, chunk_size, num_chunks):

2431 """Prefill ``req`` one ``chunk_size`` chunk per scheduler step, yielding the

2447 - boundary-state hand-offs offered (and claimed) at each step.

2448 - """

2432 + boundary-state hand-offs offered (and claimed) at each step."""

2433 computed_blocks, num_computed, _ = manager.get_computed_blocks(req)

2434 for step in range(num_chunks):

2435 manager.new_step_starts()

@@ -2509,8 +2493,7 @@ def test_boundary_states_offered_past_prompt_for_resumed_prefill():

2493 ``num_prompt_tokens``. A resumed request re-prefills its generated tokens

2494 and every group re-saves them, so filtering on the original prompt length

2495 would silently strip the mamba key for boundaries full attention still

2512 - stores; only the connector knows where its save window ends.

2513 - """

2496 + stores; only the connector knows where its save window ends."""

2497 block_size = 8

2498 prompt_len = block_size

2499 manager = make_kv_cache_manager(

tests/v1/core/test_async_scheduler.py

+2 -4

@@ -552,8 +552,7 @@ def _create_async_pp_scheduler(

552

553 def _assert_ordered_subset(delivered: list[int], emitted: list[int]) -> None:

554 """Delivered tokens must be an order-preserving subset of the emitted

555 - tokens with no duplicates (tokens are globally unique).

556 - """

555 + tokens with no duplicates (tokens are globally unique)."""

556 it = iter(emitted)

557 for token in delivered:

558 assert token in it, f"token {token} delivered out of order or twice"

@@ -563,8 +562,7 @@ def _assert_positions_consistent(req, engine: PipelinedEngine) -> None:

562 """The i-th delivered output token must be one the runner sampled for

563 exactly sequence position prompt_len + i: catches a preempted request's

564 stale output landing on a position the resumed request resampled (or

566 - vice versa), which token-stream equality alone cannot see.

567 - """

565 + vice versa), which token-stream equality alone cannot see."""

566 for i, token in enumerate(req.output_token_ids):

567 expected = req.num_prompt_tokens + i

568 actual = engine.emitted_position[token]

tests/v1/core/test_contiguous_kv_packing.py

+6 -12

@@ -178,8 +178,7 @@ def _shared_layout_config():

178 class TestCSALinearGrouping:

179 """A CSA + linear-attention model (sparse attention with a compressor ring,

180 plus sharded GDN and one TP-replicated PLE state) goes through the generic

181 - packed-group path; no model-specific branch is involved.

182 - """

181 + packed-group path; no model-specific branch is involved."""

182

183 @staticmethod

184 def _mamba_groups(groups):

@@ -194,8 +193,7 @@ class TestCSALinearGrouping:

193 def test_replicated_state_gets_its_own_group(self):

194 """The PLE state is TP-replicated and a different size, so it must not

195 share a manager group with the sharded GDN states (the NIXL worker

197 - requires exactly one single-layer replicated group).

198 - """

196 + requires exactly one single-layer replicated group)."""

197 groups = get_kv_cache_groups(_shared_layout_config(), _make_csa_linear_specs())

198

199 mamba_groups = self._mamba_groups(groups)

@@ -248,8 +246,7 @@ class TestCSALinearGrouping:

246 def test_scratch_group_survives_computed_block_truncation(self):

247 """The scratch group contributes no computed blocks, so truncating a

248 lookup result must skip it: its block size is the ring capacity, which

251 - neither divides the hit length nor bounds the (empty) block list.

252 - """

249 + neither divides the hit length nor bounds the (empty) block list."""

250 config = _shared_layout_config()

251 config.cache_config.enable_prefix_caching = True

252 groups = get_kv_cache_groups(config, _make_csa_linear_specs())

@@ -363,8 +360,7 @@ class TestCSALinearGrouping:

360 per page size, which is emitted whole, or a single-layer state that

361 cannot be split at all -- the mamba layers must be sized against it.

362 Sizing them against a narrower bucket splits them past what the block

366 - already fits, spending a pool block per extra group for no saving.

367 - """

363 + already fits, spending a pool block per extra group for no saving."""

364 config = _mock_vllm_config("BLNHC")

365 config.speculative_config = None

366 specs = {}

@@ -409,8 +405,7 @@ class TestSlidingWindowBucketCap:

405 unbalanced 8-layer paged MLA bucket. Left whole, the SWA bucket would

406 set the block width and the MLA group would fill under a third of

407 every block; capped, it splits into groups no wider than the MLA

412 - page.

413 - """

408 + page."""

409 config = _shared_layout_config()

410 specs: dict[str, KVCacheSpec] = {}

411 for layer, ratio in ((2, 2), (8, 2), (14, 2), (20, 1)):

@@ -611,8 +606,7 @@ class TestCompressorRingGroup:

606 """DeepSeek-V4.1 compressor state: a CircularBufferSpec ring (capacity

607 8) beside 128-token paged MLA groups and 64-token SWA groups. The ring

608 claims one block per request, never hashes, and does not disturb the

614 - scheduler/hash block sizes the paged groups imply.

615 - """

609 + scheduler/hash block sizes the paged groups imply."""

610 config = _shared_layout_config()

611 config.cache_config.enable_prefix_caching = True

612 specs: dict[str, KVCacheSpec] = {

tests/v1/core/test_encoder_cache_manager.py

+3 -6

@@ -367,8 +367,7 @@ def test_free_request_with_duplicate_mm_hashes():

367 """Freeing a request whose two inputs share the same mm_hash must fully

368 clean up request_cached_ids. After the first free_encoder_input call,

369 cached[mm_hash] becomes empty; the second call must still remove the

370 - remaining input_id from request_cached_ids.

371 - """

370 + remaining input_id from request_cached_ids."""

371 manager = EncoderCacheManager(cache_size=20)

372

373 req = MockRequest("r1", ["imgA", "imgA"], [4, 4])

@@ -384,8 +383,7 @@ def test_free_request_with_duplicate_mm_hashes():

383

384 def test_duplicate_mm_hash_stays_referenced_until_last_free():

385 """`cached` holds one reference per request, not per position, so freeing

387 - the first of two occurrences must not release the entry.

388 - """

386 + the first of two occurrences must not release the entry."""

387 manager = EncoderCacheManager(cache_size=20)

388 req = MockRequest("r1", ["imgA", "imgA"], [4, 4])

389

@@ -406,8 +404,7 @@ def test_duplicate_mm_hash_stays_referenced_until_last_free():

404 def test_duplicate_mm_hash_is_not_evicted_before_last_use():

405 """An item repeated within one request (an image carried across

406 conversation turns) must not become reclaimable capacity while a later

409 - occurrence still has to be spliced in, or the encoder recomputes it.

410 - """

407 + occurrence still has to be spliced in, or the encoder recomputes it."""

408 manager = EncoderCacheManager(cache_size=8)

409 req = MockRequest("r1", ["imgA", "imgA"], [4, 4])

410 other = MockRequest("r2", ["imgB"], [8])

tests/v1/core/test_kv_cache_utils.py

+10 -20

@@ -2593,8 +2593,7 @@ def test_get_kv_cache_config_kpool_tail_coowns_indexer_tensor():

2593 """The kpool tail parasitizes the indexer tensors instead of getting its

2594 own: sibling idx/tail tensors paired by layer order, zero standalone tail

2595 bytes, one shared block per request, and no prefix-caching leakage from

2596 - the kpool-sized scratch group.

2597 - """

2596 + the kpool-sized scratch group."""

2597 model_config = ModelConfig(max_model_len=8192)

2598 vllm_config = VllmConfig(model_config=model_config)

2599

@@ -2665,8 +2664,7 @@ def test_get_kv_cache_config_kpool_tail_coowns_indexer_tensor():

2664

2665 def test_glm5_kpool_tail_does_not_drag_hash_block_size():

2666 """The tail's kpool-sized scratch block (4 tokens) must not constrain the

2668 - prefix-cache hash granularity: participating groups alone decide it.

2669 - """

2667 + prefix-cache hash granularity: participating groups alone decide it."""

2668 model_config = ModelConfig(max_model_len=8192)

2669 vllm_config = VllmConfig(model_config=model_config)

2670

@@ -2742,8 +2740,7 @@ def test_get_kv_cache_config_mamba_hybrid_sharing_infeasible_no_indexer():

2740

2741 def test_get_kv_cache_config_mamba_hybrid_sharing_prepadded_mamba():

2742 """Platform-prepadded mamba pages (mamba_page_size_padded hint) must not

2745 - disable slot sharing; the layout re-pads them to the MLA page.

2746 - """

2743 + disable slot sharing; the layout re-pads them to the MLA page."""

2744 model_config = ModelConfig(max_model_len=8192)

2745 vllm_config = VllmConfig(model_config=model_config)

2746

@@ -2774,8 +2771,7 @@ def test_get_kv_cache_config_mamba_hybrid_sharing_prepadded_mamba():

2771 def test_get_kv_cache_config_mamba_hybrid_sharing_pp_balanced_projection():

2772 """Round-robin mamba grouping keeps practical PP splits balanced: every

2773 stage's largest projected mamba group slice fits its projected MLA

2777 - layers, so per-stage slot tensors all have an MLA owner.

2778 - """

2774 + layers, so per-stage slot tensors all have an MLA owner."""

2775 model_config = ModelConfig(max_model_len=8192)

2776 vllm_config = VllmConfig(model_config=model_config)

2777

@@ -2819,8 +2815,7 @@ def test_get_kv_cache_config_mamba_hybrid_sharing_pp_group_count_bump(monkeypatc

2815 """PP=4's default partition [11,11,12,11] leaves stage 0 with 2 MLA

2816 layers but a round-robin slice of 3 under the minimum 4 mamba groups;

2817 grouping bumps to 5 groups so every stage's projection keeps sharing

2822 - on instead of silently falling back on stage 0.

2823 - """

2818 + on instead of silently falling back on stage 0."""

2819 from vllm.config import ParallelConfig

2820 from vllm.distributed.utils import get_pp_indices

2821

@@ -2869,8 +2864,7 @@ def test_get_kv_cache_config_mamba_hybrid_sharing_pp_starved_stage(monkeypatch):

2864 def test_get_kv_cache_config_mamba_hybrid_sharing_beats_cross_layers_flag():

2865 """Hybrid slot sharing must take precedence over the experimental

2866 enable_cross_layers_blocks packed layout: generic packing would give the

2872 - MLA slots a strided view, breaking the contiguous virtual split.

2873 - """

2867 + MLA slots a strided view, breaking the contiguous virtual split."""

2868 model_config = ModelConfig(max_model_len=8192)

2869 kv_transfer_config = KVTransferConfig(

2870 kv_connector="NixlConnector",

@@ -3172,8 +3166,7 @@ def test_multi_run_layer_compact_strides_place_hoisted_heads():

3166 """A layer-compact run region is its own dense allocation: under LHBNC the head

3167 groups sit between the layers and the blocks, so a run's block stride is one head

3168 group's slice, not the whole page (regression test for setStorage out-of-bounds

3175 - on ROCm hybrid models).

3176 - """

3169 + on ROCm hybrid models)."""

3170 full = new_kv_cache_spec()

3171 swa = new_sliding_window_spec(sliding_window=full.block_size * 2)

3172 page = full.page_size_bytes

@@ -3252,8 +3245,7 @@ def test_mla_with_incompatible_swa_uses_one_full_allocation_group(caplog_vllm):

3245 def test_hidden_states_with_tp_scales_page_size():

3246 """When TP shrinks KV pages below the hidden-state per-token cost,

3247 get_kv_cache_groups must scale up target block sizes so that the

3255 - common page accommodates the unsharded hidden states.

3256 - """

3248 + common page accommodates the unsharded hidden states."""

3249 # Simulate TP=4 sharding a model with 8 KV heads → 2 per rank.

3250 # KV page = block_size(16) * num_kv_heads(2) * head_size(64) * dtype(2)

3251 # = 16 * 2 * 64 * 2 = 4096 bytes.

@@ -4083,8 +4075,7 @@ def test_check_enough_kv_cache_memory_reserves_null_block():

4075

4076 def test_is_full_attention_spec_unwraps_uniform_type_specs():

4077 """``UniformTypeKVCacheSpecs`` is not a ``FullAttentionSpec``, so callers

4086 - scanning groups with a bare isinstance miss DeepSeek-V4-shaped configs.

4087 - """

4078 + scanning groups with a bare isinstance miss DeepSeek-V4-shaped configs."""

4079 common = dict(num_kv_heads=1, head_size=1, dtype=torch.float32)

4080 full = FullAttentionSpec(block_size=4, **common)

4081 mla = MLAAttentionSpec(block_size=4, **common)

@@ -4253,8 +4244,7 @@ def test_no_warning_when_draft_group_is_identified(caplog_vllm):

4244

4245 def _deepseek_v4_specs(model_version="deepseek_v4"):

4246 """DeepseekV4-shaped specs: full MLA layers plus sliding-window MLA layers

4256 - at differing page sizes, with the MTP draft layer registered last.

4257 - """

4247 + at differing page sizes, with the MTP draft layer registered last."""

4248 return {

4249 "model.layers.0.self_attn.attn": new_mla_spec(),

4250 "model.layers.1.self_attn.attn": new_mla_spec(),

tests/v1/core/test_output.py

+3 -6

@@ -55,8 +55,7 @@ def _mm_feature(offset: int, length: int) -> MultiModalFeatureSpec:

55

56 def test_strip_covered_mm_data() -> None:

57 """Items fully inside the computed prefix lose their data; items touching

58 - the uncomputed region keep it; already-None data stays None.

59 - """

58 + the uncomputed region keep it; already-None data stays None."""

59 from dataclasses import replace

60

61 covered = _mm_feature(offset=0, length=100)

@@ -108,8 +107,7 @@ def _mm_feature_mixed(offset: int, length: int) -> MultiModalFeatureSpec:

107

108 def test_strip_covered_mm_data_mrope() -> None:

109 """For M-RoPE models, covered items keep their keep_on_cpu metadata fields

111 - (the worker needs them to compute positions); payload fields are dropped.

112 - """

110 + (the worker needs them to compute positions); payload fields are dropped."""

111 covered = _mm_feature_mixed(offset=0, length=100)

112 uncovered = _mm_feature_mixed(offset=300, length=100)

113

@@ -129,8 +127,7 @@ def test_strip_covered_mm_data_shm_address_item() -> None:

127 """SHM address descriptors must survive stripping even when covered: the

128 worker needs the address to resolve the payload and to acknowledge the

129 sender's writer reference. Stripping it crashed the SHM receiver cache on

132 - the second identical request (vllm-project/vllm#54994).

133 - """

130 + the second identical request (vllm-project/vllm#54994)."""

131 address_item = MultiModalKwargsItem(

132 {

133 "address": MultiModalFieldElem(data=4096, field=MultiModalBatchedField()),

tests/v1/core/test_prefix_caching.py

+25 -50

@@ -126,8 +126,7 @@ def make_kv_cache_manager(kv_cache_config: KVCacheConfig, **kwargs) -> KVCacheMa

126 """Build a ``KVCacheManager``, deriving ``scheduler_block_size`` from the

127 config (LCM of group block sizes) unless explicitly provided. This mirrors

128 ``resolve_kv_cache_block_sizes`` for the non-context-parallel case used by

129 - these tests, so callers don't have to pass it at every site.

130 - """

129 + these tests, so callers don't have to pass it at every site."""

130 kwargs.setdefault(

131 "scheduler_block_size",

132 lcm(*(g.kv_cache_spec.block_size for g in kv_cache_config.kv_cache_groups)),

@@ -3330,8 +3329,7 @@ def test_kv_cache_events_with_lora(blocks_to_cache: int):

3329 @pytest.mark.parametrize("group_id", [0, 1, 2])

3330 def test_block_stored_event_group_idx(group_id: int):

3331 """Test BlockStored events emitted by cache_full_blocks carry the correct

3333 - group_idx.

3334 - """

3332 + group_idx."""

3333 block_size = 4

3334 num_tokens = block_size * 2

3335

@@ -3624,8 +3622,7 @@ def test_block_removed_event_group_idx(group_id: int):

3622 def test_emit_cached_block_events():

3623 """emit_cached_block_events emits one BlockStored for already-cached

3624 (reused) prefix blocks, carrying the correct group_idx /

3627 - parent_block_hash / token_ids, and without mutating block state.

3628 - """

3625 + parent_block_hash / token_ids, and without mutating block state."""

3626 block_size = 4

3627 num_cached_blocks = 3

3628 kv_cache_group_id = 1

@@ -3738,8 +3735,7 @@ def test_emit_cached_block_events_zero_cached():

3735

3736 def test_eagle_enabled_removes_last_block():

3737 """Verify Eagle does NOT remove blocks when request

3741 - length is divisible by block size.

3742 - """

3738 + length is divisible by block size."""

3739 block_size = 16

3740 manager = make_kv_cache_manager(

3741 make_kv_cache_config(block_size, num_blocks=10),

@@ -4142,8 +4138,7 @@ def test_hybrid_cache_blocks_swa_tail_window_only():

4138 (its right-to-left scan stops once a contiguous match is found). Blocks

4139 earlier in the segment can never serve a hit, so

4140 ``HybridKVCacheCoordinator.cache_blocks`` should skip them rather than

4145 - polluting the prefix-cache hash map.

4146 - """

4141 + polluting the prefix-cache hash map."""

4142 block_size = 8

4143 # Full attn block_size=32, SWA block_size=8, sw=8 -> lcm=32.

4144 # tail = ceil(7/8) = 1; per_segment = 32/8 = 4.

@@ -4214,8 +4209,7 @@ def test_hybrid_cache_blocks_clamped_to_lcm():

4209 Chunks past the last lcm-aligned boundary can never participate in a

4210 cache hit (find_longest_cache_hit always returns lcm-aligned hits), so

4211 caching them only pollutes the prefix-cache hash map and keeps blocks

4217 - on the LRU list that could otherwise return to the free pool.

4218 - """

4212 + on the LRU list that could otherwise return to the free pool."""

4213 block_size = 16

4214 # Full attn block_size=32, SWA block_size=16 -> lcm=32.

4215 kv_cache_config = KVCacheConfig(

@@ -4350,8 +4344,7 @@ def test_hybrid_local_kv_retention_interval_aligns_in_manager():

4344 )

4345 def test_hybrid_local_kv_retention_interval_rejects_invalid(interval, expected_match):

4346 """A retention interval that is negative or not a multiple of

4353 - scheduler_block_size errors out at construction time.

4354 - """

4347 + scheduler_block_size errors out at construction time."""

4348 block_size = 8

4349 kv_cache_config = KVCacheConfig(

4350 num_blocks=100,

@@ -4821,8 +4814,7 @@ def test_block_lookup_cache_multi_blocks_per_key():

4814 def test_can_fit_full_sequence_swa_cap_admits_long_prompt():

4815 """Hybrid full+SWA model with a pool sized at the startup minimum should

4816 admit a prompt longer than the SWA cap, because SlidingWindowManager

4824 - recycles blocks during chunked prefill (issue #39734).

4825 - """

4817 + recycles blocks during chunked prefill (issue #39734)."""

4818 block_size = 16

4819 sliding_window = 4 * block_size # 64 tokens

4820 max_num_batched_tokens = 8 * block_size # 128 tokens

@@ -4917,8 +4909,7 @@ def test_full_sequence_admission_reserves_device_blocks_for_immediate_chunk(

4909

4910 def test_can_fit_full_sequence_full_attention_still_gates_oversized():

4911 """The cap only loosens the SWA group; a prompt that exceeds the

4920 - full-attention pool capacity must still be rejected.

4921 - """

4912 + full-attention pool capacity must still be rejected."""

4913 block_size = 16

4914 sliding_window = 4 * block_size

4915 max_num_batched_tokens = 8 * block_size

@@ -5037,8 +5028,7 @@ def _take_free_blocks(manager: KVCacheManager, num_blocks: int) -> list[KVCacheB

5028 without removing them. These ref_cnt==0 blocks stand in for evictable

5029 cache-hit blocks left behind by a previous (e.g. preempted) request, and

5030 sitting at the head guarantees a later group's external ``get_new_blocks``

5040 - would contend for them on unpatched code (issue #33775).

5041 - """

5031 + would contend for them on unpatched code (issue #33775)."""

5032 blocks: list[KVCacheBlock] = []

5033 head = manager.block_pool.free_block_queue.fake_free_list_head

5034 for _ in range(num_blocks):

@@ -5049,8 +5039,7 @@ def _take_free_blocks(manager: KVCacheManager, num_blocks: int) -> list[KVCacheB

5039

5040 def _assert_no_double_allocation(manager: KVCacheManager, req_id: str) -> None:

5041 """No physical block may be handed out twice across groups, and every

5052 - block referenced by the request must have a live ref_cnt.

5053 - """

5042 + block referenced by the request must have a live ref_cnt."""

5043 block_ids = manager.get_blocks(req_id).get_block_ids()

5044 flat = [block_id for group in block_ids for block_id in group]

5045 assert len(set(flat)) == len(flat), "Block IDs are not unique across groups"

@@ -5077,8 +5066,7 @@ def _cross_group_cache_hit(

5066 ) -> Request:

5067 """Allocate ``req_id`` with a per-group local prefix hit plus external

5068 (connector) computed tokens, driving the coordinator's two-phase path.

5080 - Returns the allocated request so callers can free it (e.g. to preempt).

5081 - """

5069 + Returns the allocated request so callers can free it (e.g. to preempt)."""

5070 block_size = _two_phase_block_size(manager)

5071 hit_blocks = _take_free_blocks(manager, num_groups * local_blocks_per_group)

5072 cache_hit = KVCacheBlocks(

@@ -5173,8 +5161,7 @@ def test_swa_free_split_keeps_cached_tail_ahead_of_scratch():

5161 """Dense retention: freeing an SWA request must place its

5162 uncached scratch blocks at the front of the free queue (recycled first)

5163 and keep its cached checkpoint blocks at the back (retained for prefix

5176 - hits). This split is always-on, independent of the retention interval.

5177 - """

5164 + hits). This split is always-on, independent of the retention interval."""

5165 block_size = 8

5166 kv_cache_config = KVCacheConfig(

5167 num_blocks=100,

@@ -5285,8 +5272,7 @@ def _make_pure_swa_manager(block_size, sliding_window, num_blocks=100, **kwargs)

5272 def test_pure_swa_retention_interval_caches_sparse_tails():

5273 """Sparse retention must work for a pure-SWA single-group model, not just

5274 hybrid models: only the per-interval tails plus the latest replay tail are

5288 - cached, and a replay still hits the latest replayable boundary.

5289 - """

5275 + cached, and a replay still hits the latest replayable boundary."""

5276 block_size = 16

5277 manager = _make_pure_swa_manager(

5278 block_size, sliding_window=block_size, retention_interval=64

@@ -5358,8 +5344,7 @@ def test_pure_swa_retention_latest_only():

5344

5345 def test_pure_swa_dense_retention_caches_all():

5346 """With retention set to ``None``, a pure-SWA model keeps dense behavior:

5361 - every block boundary is a potential hit, so all blocks are cached.

5362 - """

5347 + every block boundary is a potential hit, so all blocks are cached."""

5348 block_size = 16

5349 manager = _make_pure_swa_manager(block_size, sliding_window=block_size)

5350

@@ -5388,8 +5373,7 @@ def test_mamba_reachable_block_mask_sparsifies_retention():

5373 """Mamba state-snapshot retention: with a configured retention interval,

5374 the manager keeps one cached state per interval-sized segment (plus the

5375 latest replay boundary) instead of a snapshot per block, which is what

5391 - lets a small attention block_size avoid Mamba dominating the KV pool.

5392 - """

5376 + lets a small attention block_size avoid Mamba dominating the KV pool."""

5377 from vllm.v1.core.single_type_kv_cache_manager import MambaManager

5378

5379 block_size = 16

@@ -5427,8 +5411,7 @@ def test_mamba_reachable_block_mask_pins_shared_prefix():

5411 """A Marconi-detected shared prefix (``shared_prefix_boundary``) lands before

5412 ``num_prompt`` so the replay-boundary rule alone would drop it. The mask must

5413 pin the single state block ending on that boundary so sparse retention does

5430 - not defeat cross-request shared-prefix reuse.

5431 - """

5414 + not defeat cross-request shared-prefix reuse."""

5415 from vllm.v1.core.single_type_kv_cache_manager import MambaManager

5416

5417 block_size = 16

@@ -5476,8 +5459,7 @@ def test_mamba_shared_prefix_survives_zero_retention():

5459 detection) keeps its Mamba state block cached under

5460 ``prefix_cache_retention_interval=0``, which otherwise retains only the

5461 end-of-prompt replay boundary. Without this, a shared prefix (junction

5479 - before ``num_prompt``) would be recomputed by every sharing request.

5480 - """

5462 + before ``num_prompt``) would be recomputed by every sharing request."""

5463 block_size = 16

5464

5465 # 16-block (256-token) prompt; replay boundary is block 240 // 16 - 1 = 14.

@@ -5519,8 +5501,7 @@ def test_mamba_shared_prefix_reuse_under_zero_retention():

5501 detecting request must stay reusable by a later request under

5502 ``prefix_cache_retention_interval=0``. Without the pin the junction is

5503 masked out and the later request misses; with it (and under dense) the reuse

5522 - is preserved.

5523 - """

5504 + is preserved."""

5505 block_size = 16

5506

5507 def last_req_hit(retention, pin):

@@ -5603,8 +5584,7 @@ def test_swa_reachable_block_mask_final_partial_segment():

5584 def test_swa_reachable_block_mask_pins_shared_prefix():

5585 """SWA analog of the Mamba pin: the shared-prefix junction must keep the

5586 ``need``-block sliding-window tail ending on that boundary (not a single

5606 - block), so a windowed hit can land there under sparse retention.

5607 - """

5587 + block), so a windowed hit can land there under sparse retention."""

5588 from vllm.v1.core.single_type_kv_cache_manager import SlidingWindowManager

5589

5590 block_size = 16

@@ -5649,8 +5629,7 @@ def test_swa_reachable_block_mask_pins_shared_prefix():

5629 def test_swa_reachable_block_mask_with_dcp_scaling():

5630 """DCP shards each block's KV across ranks, scaling the effective block size.

5631 Verify that dcp_world_size > 1 scales block size in reachability calculations,

5652 - producing different masks than dcp_world_size=1.

5653 - """

5632 + producing different masks than dcp_world_size=1."""

5633 from vllm.v1.core.single_type_kv_cache_manager import SlidingWindowManager

5634

5635 block_size = 16

@@ -5709,8 +5688,7 @@ def test_swa_reachable_block_mask_sub_block_alignment_is_dense(

5688 multiple of the DCP-scaled block size) cannot be represented exactly. This

5689 happens for hybrid offloading (e.g. Gemma), where alignment_tokens is the

5690 full-attention chunk size and need not divide the SWA block size. The mask

5712 - must fall back to dense (None) rather than raise.

5713 - """

5691 + must fall back to dense (None) rather than raise."""

5692 from vllm.v1.core.single_type_kv_cache_manager import SlidingWindowManager

5693

5694 spec = SlidingWindowSpec(

@@ -5738,8 +5716,7 @@ def test_mamba_reachable_block_mask_ignores_dcp():

5716 """Mamba uses TP, not DCP: each rank holds the full recurrent state, so a

5717 state block spans kv_cache_spec.block_size tokens regardless of DCP. The

5718 mask must not scale by dcp_world_size, so retention granularity is

5741 - identical for any DCP world size.

5742 - """

5719 + identical for any DCP world size."""

5720 from vllm.v1.core.single_type_kv_cache_manager import MambaManager

5721

5722 spec = MambaSpec(

@@ -5773,8 +5750,7 @@ def test_mamba_reachable_block_mask_ignores_dcp():

5750

5751 def test_mamba_reachable_block_mask_large_dcp_stays_sparse():

5752 """A large DCP world size must not scale the Mamba block size, so it can

5776 - never collapse sparse retention into dense caching.

5777 - """

5753 + never collapse sparse retention into dense caching."""

5754 from vllm.v1.core.single_type_kv_cache_manager import MambaManager

5755

5756 spec = MambaSpec(

@@ -5809,8 +5785,7 @@ def test_swa_shared_prefix_reuse_under_zero_retention():

5785 """SWA cross-request analog: a partial shared prefix's sliding-window tail

5786 must stay reusable under ``prefix_cache_retention_interval=0``. Without

5787 the pin the junction window is masked out and a later request misses; with

5812 - it (and under dense) reuse is preserved.

5813 - """

5788 + it (and under dense) reuse is preserved."""

5789 block_size = 16

5790

5791 def last_req_hit(retention, pin):

tests/v1/core/test_scheduler.py

+28 -56

@@ -538,8 +538,7 @@ def test_throttle_defers_inflight_prefill_chunk():

538 """DP prefill balancing throttles ALL prefill compute on a throttled step,

539 not just new admissions: an in-progress (chunked) prefill already in the

540 running queue is also deferred, so the step runs decode-only, while a

541 - separate decode keeps being scheduled.

542 - """

541 + separate decode keeps being scheduled."""

542 scheduler = create_scheduler(

543 max_num_seqs=16, max_num_batched_tokens=50, enable_chunked_prefill=True

544 )

@@ -593,8 +592,7 @@ def test_throttle_capacity_bound_guard_admits():

592 """Saturation guard: if a cadence-aligned release step cannot drain the

593 waiting prefill queue (it ran out of token budget), the throttle backs off on

594 the next step so the backlog cannot grow into a TTFT avalanche -- prefills are

596 - admitted even though throttle_prefills is set.

597 - """

595 + admitted even though throttle_prefills is set."""

596 scheduler = create_scheduler(

597 max_num_seqs=16, max_num_batched_tokens=200, enable_chunked_prefill=True

598 )

@@ -1293,8 +1291,7 @@ def test_pending_hisparse_reclamation_defers_preemption(monkeypatch):

1291

1292 def test_prefix_cache_query_not_inflated_by_connector_defer():

1293 """The GPU prefix-cache query is recorded at admission, so a request the

1296 - connector defers several times is counted once, not once per retry.

1297 - """

1294 + connector defers several times is counted once, not once per retry."""

1295 num_defers_before_matching = 3

1296 scheduler = create_scheduler(

1297 enable_prefix_caching=True,

@@ -1322,8 +1319,7 @@ def test_prefix_cache_query_not_inflated_by_connector_defer():

1319

1320 def test_preemption_re_records_prefix_cache_query():

1321 """A preempted request re-enters the lookup on resume, so its recomputation

1325 - is counted again into the preempted stats.

1326 - """

1322 + is counted again into the preempted stats."""

1323 scheduler = create_scheduler(enable_prefix_caching=True)

1324 request = create_requests(num_requests=1)[0]

1325 scheduler.add_request(request)

@@ -1359,8 +1355,7 @@ def test_preemption_re_records_prefix_cache_query():

1355

1356 def test_prefix_cache_stats_not_recorded_when_caching_disabled():

1357 """With prefix caching off there is no local lookup, so admitting a request

1362 - records no phantom miss.

1363 - """

1358 + records no phantom miss."""

1359 scheduler = create_scheduler(enable_prefix_caching=False)

1360 for request in create_requests(num_requests=2):

1361 scheduler.add_request(request)

@@ -1374,8 +1369,7 @@ def test_prefix_cache_stats_not_recorded_when_caching_disabled():

1369

1370 def test_prefix_cache_stats_counted_once_for_retried_then_scheduled_request():

1371 """A real cache hit rejected once by allocate_slots and admitted on the next

1377 - step is counted exactly once, hits included.

1378 - """

1372 + step is counted exactly once, hits included."""

1373 block_size = 16

1374 scheduler = create_scheduler(

1375 enable_prefix_caching=True,

@@ -3094,8 +3088,7 @@ def create_requests_with_priority(

3088

3089 def test_priority_scheduling_basic_ordering():

3090 """Test that requests are scheduled in priority order

3097 - (lower value = higher priority).

3098 - """

3091 + (lower value = higher priority)."""

3092 scheduler = create_scheduler_with_priority()

3093

3094 # Create requests with different priorities

@@ -3124,8 +3117,7 @@ def test_priority_scheduling_basic_ordering():

3117

3118 def test_priority_scheduling_arrival_time_tiebreaker():

3119 """Test that arrival time is used

3127 - as tiebreaker when priorities are equal.

3128 - """

3120 + as tiebreaker when priorities are equal."""

3121 scheduler = create_scheduler_with_priority()

3122

3123 # Create requests with same priority but different arrival times

@@ -3282,8 +3274,7 @@ def test_priority_scheduling_preemption():

3274

3275 def test_priority_scheduling_no_preemption_when_space_available():

3276 """Test that preemption doesn't happen

3285 - when there's space for new requests.

3286 - """

3277 + when there's space for new requests."""

3278 scheduler = create_scheduler_with_priority(

3279 max_num_seqs=3, # Allow 3 concurrent requests

3280 max_num_batched_tokens=200, # Sufficient token budget

@@ -3336,8 +3327,7 @@ def test_priority_scheduling_no_preemption_when_space_available():

3327

3328 def test_priority_scheduling_preemption_victim_selection():

3329 """Test that the correct victim is selected for

3339 - preemption based on priority and arrival time.

3340 - """

3330 + preemption based on priority and arrival time."""

3331 # This test verifies the priority-based victim selection logic

3332 # by checking the waiting queue order after adding requests with different

3333 # priorities

@@ -3457,8 +3447,7 @@ def test_priority_scheduling_waiting_queue_order():

3447

3448 def test_priority_scheduling_fcfs_fallback():

3449 """Test that FCFS behavior is maintained when all

3460 - requests have same priority.

3461 - """

3450 + requests have same priority."""

3451 scheduler = create_scheduler_with_priority()

3452

3453 # Create requests with same priority but different arrival times

@@ -3527,8 +3516,7 @@ def test_priority_scheduling_with_limited_slots():

3516

3517 def test_priority_scheduling_heap_property():

3518 """Test that the waiting queue maintains heap

3530 - property for priority scheduling.

3531 - """

3519 + property for priority scheduling."""

3520 scheduler = create_scheduler_with_priority(

3521 max_num_seqs=1, # Only one request can run at a time

3522 )

@@ -3753,8 +3741,7 @@ def test_priority_scheduling_preemption_and_resumption_when_out_of_kv(

3741 use_ec_connector, ec_role, use_v2_model_runner

3742 ):

3743 """Test that priority scheduling preempts lower priority requests

3756 - when out of KV cache space.

3757 - """

3744 + when out of KV cache space."""

3745 # Create scheduler with very limited memory to force preemption

3746 scheduler = create_scheduler_with_priority(

3747 max_num_seqs=2, # Allow multiple requests

@@ -3899,8 +3886,7 @@ def test_chunked_prefill_disabled_for_encoder_decoder(

3886 enable_chunked_prefill: bool, is_encoder_decoder: bool, expect_enabled: bool

3887 ) -> None:

3888 """Validate that chunked prefill is appropriately disabled for

3902 - encoder-decoder models.

3903 - """

3889 + encoder-decoder models."""

3890 scheduler_config = SchedulerConfig(

3891 enable_chunked_prefill=enable_chunked_prefill,

3892 is_encoder_decoder=is_encoder_decoder,

@@ -3929,8 +3915,7 @@ def _validate_chunked_prefill_settings_for_encoder_decoder(

3915 scheduler_config: SchedulerConfig, is_encoder_decoder: bool, expect_enabled: bool

3916 ) -> None:

3917 """Validate chunked prefill settings in the scheduler config for

3932 - encoder-decoder models.

3933 - """

3918 + encoder-decoder models."""

3919 assert scheduler_config.enable_chunked_prefill is expect_enabled

3920 if is_encoder_decoder:

3921 # Encoder-decoder models should automatically disable chunked multimodal

@@ -4020,8 +4005,7 @@ def _assert_right_encoder_inputs(

4005 ):

4006 """Verify that requests/mm_hashes should (not) in scheduled encoder input

4007 If check_exist is False, this function returns True

4023 - if requests are NOT in encoder inputs

4024 - """

4008 + if requests are NOT in encoder inputs"""

4009 # Get the scheduled encoder inputs

4010 # NOTE: scheduled_encoder_inputs is a dictionary with request id as key

4011 scheduled_encoder_inputs = output.scheduled_encoder_inputs

@@ -4182,8 +4166,7 @@ def test_ec_connector_text_only_request(use_kv_connector):

4166 @pytest.mark.parametrize("use_kv_connector", [False, True])

4167 def test_ec_connector_cache_hit_external_load(use_kv_connector):

4168 """Test ec_consumer loads from external cache when hit.

4185 - A normal basic operation for EPD disaggrgation

4186 - """

4169 + A normal basic operation for EPD disaggrgation"""

4170 scheduler = create_scheduler(

4171 model="llava-hf/llava-1.5-7b-hf",

4172 enable_prefix_caching=True,

@@ -4630,8 +4613,7 @@ def test_priority_scheduling_ec_connector_preemption_and_resumption(

4613 cache_exist, use_kv_connector

4614 ):

4615 """Test that priority scheduling preempts lower priority requests

4633 - when out of KV cache space.

4634 - """

4616 + when out of KV cache space."""

4617 # Create scheduler with very limited memory to force preemption

4618 scheduler = create_scheduler_with_priority(

4619 model="llava-hf/llava-1.5-7b-hf",

@@ -5300,8 +5282,7 @@ def test_scheduler_kv_connector_stats():

5282

5283 def test_ec_connector_update_connector_output_called():

5284 """Test that worker-side EC connector output is forwarded to the

5303 - EC connector's update_connector_output hook.

5304 - """

5285 + EC connector's update_connector_output hook."""

5286 scheduler = create_scheduler(

5287 model="llava-hf/llava-1.5-7b-hf",

5288 use_ec_connector=True,

@@ -5506,8 +5487,7 @@ def test_variable_length_cross_attn_block_allocation():

5487

5488 def test_cross_attn_blocks_not_over_allocated():

5489 """Test that cross-attention blocks are not over-allocated compared to

5509 - what each request actually needs.

5510 - """

5490 + what each request actually needs."""

5491 from math import ceil

5492

5493 block_size = 16

@@ -5603,8 +5583,7 @@ def test_cross_attn_blocks_not_under_allocated():

5583

5584 def test_cross_attn_zero_blocks_without_encoder_inputs():

5585 """Test that requests without encoder inputs get zero cross-attention

5606 - blocks, even when the scheduler is configured for encoder-decoder.

5607 - """

5586 + blocks, even when the scheduler is configured for encoder-decoder."""

5587 block_size = 16

5588 scheduler = _create_encoder_decoder_scheduler(block_size=block_size)

5589

@@ -5709,8 +5688,7 @@ def test_free_encoder_inputs_respects_unconfirmed_placeholders():

5688 evicted entry and crashes the engine with "Encoder cache miss". The

5689 scheduler must retain the input until the *confirmed* progress

5690 (num_computed_tokens - num_output_placeholders) passes the range end, so

5712 - that no pending rejection can rewind into the range.

5713 - """

5691 + that no pending rejection can rewind into the range."""

5692 scheduler = create_scheduler(

5693 model="llava-hf/llava-1.5-7b-hf",

5694 num_speculative_tokens=3,

@@ -5821,8 +5799,7 @@ def test_free_encoder_inputs_defers_for_eagle_lookahead():

5799 position so the drafter's +1 look-ahead mm-embedding gather (which reads one

5800 position past the target's computed range) still finds it cached. This is

5801 the primary mechanism that prevents the drafter "Encoder cache miss"; the

5824 - worker-side token-embedding fallback is only a backstop.

5825 - """

5802 + worker-side token-embedding fallback is only a backstop."""

5803 scheduler = create_scheduler(model="llava-hf/llava-1.5-7b-hf")

5804 # create_scheduler only builds ngram spec configs; force the eagle path that

5805 # _free_encoder_inputs keys off (its read-ahead deferral).

@@ -5852,8 +5829,7 @@ def test_free_encoder_inputs_defers_for_eagle_lookahead():

5829

5830 def test_free_encoder_inputs_unchanged_without_spec_decode():

5831 """Without speculative decoding, encoder inputs are freed as soon as

5855 - num_computed_tokens passes the placeholder range, as before.

5856 - """

5832 + num_computed_tokens passes the placeholder range, as before."""

5833 scheduler = create_scheduler(model="llava-hf/llava-1.5-7b-hf")

5834 mm_positions = [[PlaceholderRange(offset=50, length=100)]]

5835 request = create_requests(

@@ -5884,8 +5860,7 @@ def test_encoder_cache_retained_across_preemption_and_resume():

5860 encoder input must pull the still-cached entry back out of `freeable`

5861 without scheduling a recompute, keeping the scheduler and worker

5862 consistent. The spec-rollback retention margin does not gate this path,

5887 - so it is covered separately here.

5888 - """

5863 + so it is covered separately here."""

5864 scheduler = create_scheduler(

5865 model="llava-hf/llava-1.5-7b-hf",

5866 num_speculative_tokens=3,

@@ -5929,8 +5904,7 @@ def test_encoder_cache_recomputed_when_evicted_during_preemption():

5904 pressure before it resumes, the scheduler reports the mm_hash as freed

5905 (so the worker drops it) and a resume must schedule a recompute rather

5906 than assume the worker still holds it. check_and_update_cache must

5932 - return False so the encoder input is re-scheduled.

5933 - """

5907 + return False so the encoder input is re-scheduled."""

5908 scheduler = create_scheduler(

5909 model="llava-hf/llava-1.5-7b-hf",

5910 num_speculative_tokens=3,

@@ -6581,8 +6555,7 @@ def _decode_ready_request(scheduler):

6555

6556 def test_update_draft_token_ids_strips_ngram_padding():

6557 """ngram_gpu pads unfilled draft slots with -1; manager validation must

6584 - strip them before grammar.validate_tokens (which otherwise raises).

6585 - """

6558 + strip them before grammar.validate_tokens (which otherwise raises)."""

6559 scheduler = create_scheduler(num_speculative_tokens=4)

6560 request = _decode_ready_request(scheduler)

6561

@@ -6601,8 +6574,7 @@ def test_update_draft_token_ids_strips_ngram_padding():

6574

6575 def test_update_draft_token_ids_in_output_strips_padding():

6576 """Same guard on the output path; the -1 pad-back for the rejected count

6604 - is preserved (only the input to manager validation is stripped).

6605 - """

6577 + is preserved (only the input to manager validation is stripped)."""

6578 scheduler = create_scheduler(num_speculative_tokens=4)

6579 request = _decode_ready_request(scheduler)

6580

tests/v1/core/test_swa_inflight_window_free.py

+4 -8

@@ -97,8 +97,7 @@ def test_num_in_flight_tokens_accounting():

97

98 def test_swa_free_waits_for_in_flight_step():

99 """Async: out-of-window blocks stay allocated until the step that still

100 - reads them has been processed.

101 - """

100 + reads them has been processed."""

101 scheduler = _create_swa_scheduler(async_scheduling=True)

102 request = create_requests(

103 num_requests=1, num_tokens=NUM_PROMPT_TOKENS, block_size=BLOCK_SIZE

@@ -133,8 +132,7 @@ def test_swa_free_waits_for_in_flight_step():

132

133 def test_swa_free_immediate_when_sync():

134 """Sync: no in-flight step at schedule time, frees happen at the first

136 - decode allocation as before.

137 - """

135 + decode allocation as before."""

136 scheduler = _create_swa_scheduler(async_scheduling=False)

137 request = create_requests(

138 num_requests=1, num_tokens=NUM_PROMPT_TOKENS, block_size=BLOCK_SIZE

@@ -173,8 +171,7 @@ def test_swa_admission_cap_accounts_for_overlapping_batches():

171 def test_chunked_local_free_waits_for_in_flight_step():

172 """Chunked-local attention frees whole chunks left of the current one, and

173 is exposed to the same load-WAR: with async scheduling those chunks must

176 - stay allocated until the in-flight step that still reads them settles.

177 - """

174 + stay allocated until the in-flight step that still reads them settles."""

175 scheduler = _create_chunked_scheduler(async_scheduling=True)

176 request = create_requests(

177 num_requests=1, num_tokens=NUM_PROMPT_TOKENS, block_size=BLOCK_SIZE

@@ -229,8 +226,7 @@ def test_connector_finish_frees_on_settled_basis():

226 is handed to a KV connector (simple CPU offload / NIXL store), must use the

227 same processed-token basis. Otherwise the connector reads/hands off a block

228 the still-in-flight step is optimistically counted as done with, which is

232 - the load-path WAR this fix closes.

233 - """

229 + the load-path WAR this fix closes."""

230 scheduler = create_scheduler(

231 block_size=BLOCK_SIZE,

232 async_scheduling=True,

tests/v1/cudagraph/test_breakable_cudagraph.py

+3 -6

@@ -80,8 +80,7 @@ def test_piecewise_capture_builds_fresh_metadata_for_both_passes():

80 @pytest.fixture(autouse=True)

81 def _reset_breakable_tls():

82 """Defensively clear thread-local capture state between tests so a

83 - failure in one test can't leak "nested capture" errors into the next.

84 - """

83 + failure in one test can't leak "nested capture" errors into the next."""

84 from vllm.compilation.breakable_cudagraph import BreakableCUDAGraphCapture

85

86 BreakableCUDAGraphCapture._tls.active = None

@@ -291,8 +290,7 @@ def test_capture_replay_matches_eager_simple(cuda_capture_stream):

290 def test_decorator_breaks_when_invoked_inside_capture(cuda_capture_stream):

291 """Verify @eager_break_during_capture correctly routes through

292 add_eager when inside a capture context, and runs straight through

294 - when there's no active capture.

295 - """

293 + when there's no active capture."""

294 from vllm.compilation.breakable_cudagraph import (

295 BreakableCUDAGraphCapture,

296 eager_break_during_capture,

@@ -554,8 +552,7 @@ def _mock_vllm_config():

552

553 def test_wrapper_falls_through_on_runtime_mode_mismatch():

554 """A PIECEWISE-scoped wrapper runs other modes eagerly, capturing

557 - nothing, so an outer CUDAGraphWrapper can own them.

558 - """

555 + nothing, so an outer CUDAGraphWrapper can own them."""

556 from vllm.compilation.breakable_cudagraph import BreakableCUDAGraphWrapper

557 from vllm.config import CUDAGraphMode

558 from vllm.forward_context import BatchDescriptor, set_forward_context

tests/v1/determinism/test_rms_norm_batch_invariant.py

+1 -2

@@ -275,8 +275,7 @@ def test_rms_norm_per_block_quant_batch_invariant(

275 hidden_size: int, dtype: torch.dtype, seed: int

276 ):

277 """C++ ``rms_norm_per_block_quant`` must be batch invariant across the

278 - block threshold (compiled fp8 block-quant path; block pinned to 512).

279 - """

278 + block threshold (compiled fp8 block-quant path; block pinned to 512)."""

279 import vllm._custom_ops as ops

280

281 device = torch.device(DEVICE_TYPE)

tests/v1/distributed/test_async_llm_dp.py

+4 -8

@@ -547,8 +547,7 @@ async def test_dp_pause_keep_then_resume(expert_parallel: bool):

547 async def test_dp_pause_keep_race_staggered_engines():

548 """Race: send pause(keep) to engine 0, then add two requests,

549 then pause(keep) to engine 1. Ensures no deadlock when pause

550 - requests are staggered and requests arrive in between.

551 - """

550 + requests are staggered and requests arrive in between."""

551 if DP_SIZE != 2:

552 pytest.skip("test_dp_pause_keep_race_staggered_engines requires DP_SIZE=2")

553

@@ -718,8 +717,7 @@ async def test_dp_pause_barrier_request_deadlock():

717 async def test_dp_pause_wait_mode_drains_in_flight():

718 """mode="wait" through the DP consensus path: the pause lets the

719 in-flight request run to completion while the consensus is pending,

721 - resolves only once drained, and the engine generates again on resume.

722 - """

720 + resolves only once drained, and the engine generates again on resume."""

721 with ExitStack() as after:

722 engine = AsyncLLM.from_engine_args(_get_dp_pause_engine_args(True))

723 after.callback(engine.shutdown)

@@ -750,8 +748,7 @@ async def test_dp_pause_wait_mode_drains_in_flight():

748 async def test_dp_pause_while_asleep():

749 """Pausing a sleeping DP engine: sleeping ranks skip dummy batches yet

750 must still reach the pause consensus, and the completion barrier must

753 - be harmless on workers whose memory is unmapped.

754 - """

751 + be harmless on workers whose memory is unmapped."""

752 with ExitStack() as after:

753 engine = AsyncLLM.from_engine_args(_get_dp_pause_engine_args(True))

754 after.callback(engine.shutdown)

@@ -787,8 +784,7 @@ async def test_dp_pause_completion_implies_device_idle(

784 """A resolved pause future promises an idle device. Inflate the dummy

785 batches the pause consensus manufactures so any work the pause fails to

786 wait on stays visible, then require a quiet worker stream the moment

790 - pause_generation() returns.

791 - """

787 + pause_generation() returns."""

788 with ExitStack() as after:

789 # The probes below ship functions through collective_rpc.

790 monkeypatch.setenv("VLLM_ALLOW_INSECURE_SERIALIZATION", "1")

tests/v1/distributed/test_external_lb_dp.py

+1 -2

@@ -24,8 +24,7 @@ TP_SIZE = int(os.getenv("TP_SIZE", "1"))

24

25 class ExternalLBServerManager:

26 """Manages data parallel vLLM server instances for external

27 - load balancer testing.

28 - """

27 + load balancer testing."""

28

29 def __init__(

30 self,

tests/v1/distributed/test_hybrid_lb_dp.py

+1 -2

@@ -30,8 +30,7 @@ DP_SIZE_LOCAL = DP_SIZE // NUM_NODES # 2 ranks per node

30 class HybridLBServerManager:

31 """Manages hybrid data parallel vLLM server instances where each node

32 runs a single logical API server that balances requests only to the

33 - DP engines running on that same node.

34 - """

33 + DP engines running on that same node."""

34

35 def __init__(

36 self,

tests/v1/distributed/test_internal_lb_dp.py

+3 -6

@@ -107,8 +107,7 @@ async def _run_request_bursts(

107

108 class MultinodeInternalLBServerManager:

109 """Manages multi-node data parallel vLLM server instances for internal

110 - load balancer testing using --headless mode.

111 - """

110 + load balancer testing using --headless mode."""

111

112 def __init__(

113 self,

@@ -242,8 +241,7 @@ class MultinodeInternalLBServerManager:

241

242 class APIOnlyServerManager:

243 """Manages API-only server (Node 0) and headless engines server (Node 1)

245 - for testing separated API server and engine configuration.

246 - """

244 + for testing separated API server and engine configuration."""

245

246 def __init__(

247 self,

@@ -645,8 +643,7 @@ async def test_api_only_multinode_dp_completion_streaming(

643 model_name: str,

644 ) -> None:

645 """Test API-only server streaming with all engines on separate

648 - headless server.

649 - """

646 + headless server."""

647 prompt = "What is an LLM?"

648

649 async def make_streaming_request():

tests/v1/distributed/test_pp_dp_v2.py

+4 -8

@@ -82,8 +82,7 @@ async def _generate(engine: AsyncLLM, prompt: str, max_tokens: int) -> int:

82 @pytest.mark.parametrize("async_scheduling", [True, False])

83 async def test_pp_dp_v2_low_concurrency(async_scheduling: bool):

84 """A single in-flight request at a time, repeated, to exercise the

85 - PP slot ring under empty batches between decodes.

86 - """

85 + PP slot ring under empty batches between decodes."""

86 with ExitStack() as after:

87 engine = AsyncLLM.from_engine_args(_engine_args(async_scheduling))

88 after.callback(engine.shutdown)

@@ -97,8 +96,7 @@ async def test_pp_dp_v2_low_concurrency(async_scheduling: bool):

96 @pytest.mark.parametrize("async_scheduling", [True, False])

97 async def test_pp_dp_v2_mid_concurrency(async_scheduling: bool):

98 """64 concurrent requests, staggered, to exercise the steady-state

100 - DP all-reduce + PP slot-ring path.

101 - """

99 + DP all-reduce + PP slot-ring path."""

100 with ExitStack() as after:

101 engine = AsyncLLM.from_engine_args(_engine_args(async_scheduling))

102 after.callback(engine.shutdown)

@@ -114,8 +112,7 @@ async def test_pp_dp_v2_mid_concurrency(async_scheduling: bool):

112 @pytest.mark.asyncio

113 async def test_pp_dp_v2_abort_mid_decode():

114 """Cancel half the in-flight requests mid-stream and confirm the

117 - engine survives the abort storm.

118 - """

115 + engine survives the abort storm."""

116 with ExitStack() as after:

117 engine = AsyncLLM.from_engine_args(_engine_args(async_scheduling=True))

118 after.callback(engine.shutdown)

@@ -155,8 +152,7 @@ async def test_pp_dp_v2_abort_mid_decode():

152 @pytest.mark.asyncio

153 async def test_pp_dp_v2_pause_resume():

154 """Pause an engine with a request in flight, then resume and confirm

158 - new requests still work.

159 - """

155 + new requests still work."""

156 with ExitStack() as after:

157 engine = AsyncLLM.from_engine_args(_engine_args(async_scheduling=True))

158 after.callback(engine.shutdown)

tests/v1/e2e/general/test_async_scheduling.py

+2 -4

@@ -48,8 +48,7 @@ def test_without_spec_decoding(

48 monkeypatch: pytest.MonkeyPatch,

49 ):

50 """Test consistency of combos of async scheduling, preemption,

51 - uni/multiproc executor, prefill chunking.

52 - """

51 + uni/multiproc executor, prefill chunking."""

52 struct_outputs = StructuredOutputsParams(json=sample_json_schema)

53 test_sampling_params: list[dict[str, Any]] = [

54 dict(),

@@ -205,8 +204,7 @@ def run_tests(

204 test_sampling_params: list[dict[str, Any]],

205 ):

206 """Test consistency of combos of async scheduling, preemption,

208 - uni/multiproc executor with spec decoding.

209 - """

207 + uni/multiproc executor with spec decoding."""

208 # Flex attention supports float32.

209 attention_config = {"backend": "FLEX_ATTENTION"}

210

tests/v1/e2e/general/test_attention_backend_per_kind.py

+1 -2

@@ -72,8 +72,7 @@ def test_backend_per_kind_splits_groups(backend_per_kind, monkeypatch):

72 )

73 def test_backend_per_kind_overrides_global_backend(monkeypatch):

74 """A per-kind entry wins over the global ``backend`` for its kind; other

75 - kinds fall back to the global backend.

76 - """

75 + kinds fall back to the global backend."""

76 monkeypatch.setenv("VLLM_ALLOW_INSECURE_SERIALIZATION", "1")

77 llm = LLM(

78 model=MODEL,

tests/v1/e2e/general/test_context_length.py

+2 -4

@@ -34,8 +34,7 @@ def test_decoder_max_context_length_validation(

34 max_tokens: int,

35 ) -> None:

36 """Check vLLM decoder model input validation for edge cases where

37 - the prompt length is (almost) equal to the max model length.

38 - """

37 + the prompt length is (almost) equal to the max model length."""

38 prompt_ids = [[43] * prompt_len]

39

40 with vllm_runner(

@@ -72,8 +71,7 @@ def test_auto_fit_max_model_len_rejects_oversized_input(

71 """When max_model_len='auto' and KV cache memory is very limited,

72 the engine auto-fits max_model_len to a small value. The frontend

73 must see this reduced value and reject prompts that exceed it,

75 - rather than accepting them and hanging.

76 - """

74 + rather than accepting them and hanging."""

75 # Use a small KV cache budget to force auto-fit to a small

76 # max_model_len. Pin block_size=16 so the budget is independent

77 # of the platform's default block size. One block for this model is

tests/v1/e2e/general/test_min_tokens.py

+1 -2

@@ -172,8 +172,7 @@ def llm_v1():

172

173 def test_min_tokens_with_terminal_structured_output(llm_v1: LLM):

174 """A grammar whose only legal continuation is EOS must stop under

175 - min_tokens instead of sampling outside the grammar.

176 - """

175 + min_tokens instead of sampling outside the grammar."""

176 params = SamplingParams(

177 temperature=GREEDY,

178 min_tokens=20,

tests/v1/e2e/general/test_sharded_sampling.py

+2 -4

@@ -7,8 +7,7 @@ slightly different kernel/collective states (measured: same-state boots are

7 bitwise identical across the two modes; cross-state boots differ by up to

8 ~0.25 logprob and can flip a near-tie token). A real sharding bug — rows

9 routed to the wrong request, misaligned gathers — produces many-nats

10 -divergence and scrambled top-k sets, far outside these bounds.

11 -"""

10 +divergence and scrambled top-k sets, far outside these bounds."""

11

12 import pytest

13

@@ -107,8 +106,7 @@ def _assert_logprob_dicts_close(ref_lps, out_lps, what: str) -> None:

106 @multi_gpu_test(num_gpus=2)

107 def test_sharded_sampling_outputs_match(monkeypatch: pytest.MonkeyPatch):

108 """Generation, logprobs, and prompt logprobs match between batch-sharded

110 - sampling and the replicated fallback, up to measured boot-state noise.

111 - """

109 + sampling and the replicated fallback, up to measured boot-state noise."""

110 monkeypatch.setenv("VLLM_USE_V2_MODEL_RUNNER", "1")

111 # Required for the collective_rpc mode check below.

112 monkeypatch.setenv("VLLM_ALLOW_INSECURE_SERIALIZATION", "1")

tests/v1/e2e/spec_decode/acceptance_rates/mtp_other/test_synthetic.py

+1 -2

@@ -11,8 +11,7 @@ from ...utils import compute_acceptance_len, get_test_prompts

11 @single_gpu_only

12 def test_synthetic_acceptance_rate(vllm_runner):

13 """Verify that synthetic rejection sampling produces an acceptance

14 - length close to the requested mean acceptance length.

15 - """

14 + length close to the requested mean acceptance length."""

15 num_spec_tokens = 3

16 expected_acceptance_len = 1.875

17 tolerance = 0.15

tests/v1/e2e/spec_decode/draft_model/test_draft_model.py

+7 -14

@@ -167,8 +167,7 @@ def test_draft_model_tensor_parallelism(vllm_runner):

167 @multi_gpu_only(num_gpus=2)

168 def test_draft_model_engine_args_tensor_parallelism():

169 """Ensure the vllm_config for the draft model is created correctly,

170 - and independently of the target model (quantization, TP, etc.)

171 - """

170 + and independently of the target model (quantization, TP, etc.)"""

171 engine_args = EngineArgs(

172 model="Qwen/Qwen3-1.7B-FP8", # <<< tgt quantized

173 tensor_parallel_size=2,

@@ -199,8 +198,7 @@ def test_draft_model_engine_args_tensor_parallelism():

198

199 def _apply_draft_moe_backend(vllm_config: VllmConfig) -> VllmConfig:

200 """Replicate SpecDecodeBaseProposer._create_draft_vllm_config logic

202 - so we can test it without instantiating a full proposer.

203 - """

201 + so we can test it without instantiating a full proposer."""

202 spec_cfg = vllm_config.speculative_config

203 if spec_cfg.moe_backend is not None:

204 return replace(

@@ -225,8 +223,7 @@ def _platform_moe_backend(monkeypatch: pytest.MonkeyPatch) -> MoEBackend:

223

224 def test_draft_model_moe_backend_override(monkeypatch: pytest.MonkeyPatch):

225 """When moe_backend is set in speculative_config, the draft VllmConfig

228 - should use it while the target keeps its own setting.

229 - """

226 + should use it while the target keeps its own setting."""

227 target_moe_backend = _platform_moe_backend(monkeypatch)

228 engine_args = EngineArgs(

229 model="Qwen/Qwen3-1.7B",

@@ -251,8 +248,7 @@ def test_draft_model_moe_backend_override(monkeypatch: pytest.MonkeyPatch):

248

249 def test_draft_model_moe_backend_inherits_target(monkeypatch: pytest.MonkeyPatch):

250 """When moe_backend is not set in speculative_config, the draft should

254 - inherit the target's moe_backend.

255 - """

251 + inherit the target's moe_backend."""

252 target_moe_backend = _platform_moe_backend(monkeypatch)

253 engine_args = EngineArgs(

254 model="Qwen/Qwen3-1.7B",

@@ -275,8 +271,7 @@ def test_draft_model_moe_backend_inherits_target(monkeypatch: pytest.MonkeyPatch

271

272 def test_draft_model_moe_backend_default_auto():

273 """When neither target nor draft set moe_backend explicitly, both should

278 - default to 'auto'.

279 - """

274 + default to 'auto'."""

275 engine_args = EngineArgs(

276 model="Qwen/Qwen3-1.7B",

277 tensor_parallel_size=1,

@@ -297,8 +292,7 @@ def test_draft_model_moe_backend_default_auto():

292

293 def test_draft_model_engine_args_rejects_invalid_tp_argname():

294 """The user should pass "draft_tensor_parallel_size" rather than

300 - "tensor_parallel_size". We enforce this with validation.

301 - """

295 + "tensor_parallel_size". We enforce this with validation."""

296 engine_args = EngineArgs(

297 model="Qwen/Qwen3-1.7B",

298 tensor_parallel_size=1,

@@ -315,8 +309,7 @@ def test_draft_model_engine_args_rejects_invalid_tp_argname():

309

310 def assert_draft_model_correctness(args: ArgsTest, vllm_runner):

311 """Compare the outputs using and not using speculative decoding.

318 - In the greedy decoding case, the outputs must match EXACTLY.

319 - """

312 + In the greedy decoding case, the outputs must match EXACTLY."""

313 test_prompts: list[Messages] = get_messages(

314 dataset=args.dataset, n=args.num_prompts

315 )

tests/v1/e2e/spec_decode/test_sharded_sampling.py

+1 -2

@@ -28,8 +28,7 @@ def test_mtp_sharded_sampling_equivalence(monkeypatch: pytest.MonkeyPatch):

28 sampling under MTP spec decoding: the collectives move the same logits

29 bytes, Gumbel keys derive from (request slot, position, seed), and slot

30 assignment is rank-deterministic. Both runs here are spec decode with

31 - identical math, so outputs must match exactly.

32 - """

31 + identical math, so outputs must match exactly."""

32 tp_size = 2

33 _skip_if_insufficient_gpus_for_tp(tp_size)

34 model_name = "Qwen/Qwen3.5-0.8B-Base"

tests/v1/e2e/test_hybrid_chunked_prefill.py

+1 -2

@@ -67,8 +67,7 @@ def test_mtp_speculative_mixed_batch_short_prefill(

67 vllm_runner, model_name, enable_prefix_caching

68 ):

69 """Test to ensure MTP speculative decoding correctly handles

70 - short prefill chunks that fall below the reorder_batch_threshold.

71 - """

70 + short prefill chunks that fall below the reorder_batch_threshold."""

71 # Set so large that both prefills will be classified as decodes in a mixed batch

72 # note, with prefix caching we require chunk_size >= mamba_block_size

73 chunk_size = 256 if not enable_prefix_caching else 2048

tests/v1/ec_connector/unit/cpu/scheduler/test_embedding_cache.py

+1 -2

@@ -381,8 +381,7 @@ def test_pin_if_ready_none_when_absent():

381

382 def test_pin_if_ready_returns_unpinned_entry_when_not_ready():

383 """The producer needs "present but not ready" told apart from "absent":

384 - the first is NACKed as retryable, the second as a miss.

385 - """

384 + the first is NACKed as retryable, the second as a miss."""

385 cache = EmbeddingCache(num_blocks=4)

386 cache.alloc("h", 1) # not-ready

387 entry = cache.pin_if_ready("h")

tests/v1/ec_connector/unit/cpu/scheduler/test_scheduler.py

+12 -24

@@ -131,8 +131,7 @@ def test_has_cache_item_false_when_not_consumer(monkeypatch):

131

132 def test_connector_keys_on_identifier_not_mm_hash(monkeypatch):

133 """The connector must key the encoder cache on feature.identifier (what

134 - has_cache_item is called with), NOT feature.mm_hash.

135 - """

134 + has_cache_item is called with), NOT feature.mm_hash."""

135 s = _make_scheduler(monkeypatch)

136 req = _Request([_Feature("PROC_KEY", length=1, identifier="ENC_KEY")])

137 s.update_state_after_alloc(req, 0)

@@ -202,8 +201,7 @@ def test_loads_only_serves_hashes_touched_this_step(monkeypatch):

201

202 def test_repeated_reload_same_step_loads_once(monkeypatch):

203 """The same mm_hash requested twice in one step must appear in

205 - meta.loads exactly once.

206 - """

204 + meta.loads exactly once."""

205 s = _make_scheduler(monkeypatch)

206 _seed_cached(s, "a", n_blocks=2)

207

@@ -226,8 +224,7 @@ def test_load_not_emitted_for_uncached_entry(monkeypatch):

224 def test_load_pin_protects_blocks_until_completion(monkeypatch):

225 """Loaded blocks stay pinned until the worker reports the load memcpy

226 complete. Observable: a save that needs eviction cannot reclaim a

229 - still-pinned loaded entry, but can once the completion report arrives.

230 - """

227 + still-pinned loaded entry, but can once the completion report arrives."""

228 # 2 blocks, both occupied by "a" (ready).

229 s = _make_scheduler(monkeypatch, num_blocks=2)

230 _seed_cached(s, "a", n_blocks=2)

@@ -254,8 +251,7 @@ def test_load_pin_protects_blocks_until_completion(monkeypatch):

251

252 def _can_evict(s: ECCPUScheduler, mm_hash: str, n_blocks: int) -> bool:

253 """Whether a new save of `mm_hash` can claim `n_blocks`, i.e. whether the

257 - blocks currently held by other entries are reclaimable.

258 - """

254 + blocks currently held by other entries are reclaimable."""

255 s.update_state_after_alloc(_Request([_Feature(mm_hash, length=n_blocks)]), 0)

256 meta = s.build_connector_meta(scheduler_output=None)

257 return mm_hash in meta.saves

@@ -315,8 +311,7 @@ def test_late_report_does_not_release_a_later_load_of_same_hash(monkeypatch):

311

312 def test_region_full_skips_save_and_never_blocks(monkeypatch):

313 """When the region is fully occupied by pinned entries, new saves are

318 - silently skipped and ensure_cache_available never blocks.

319 - """

314 + silently skipped and ensure_cache_available never blocks."""

315 s = _make_scheduler(monkeypatch, num_blocks=1)

316 _seed_cached(s, "pinned", n_blocks=1)

317 s._cache.pin("pinned")

@@ -335,8 +330,7 @@ def test_region_full_skips_save_and_never_blocks(monkeypatch):

330

331 def test_producer_only_never_emits_loads(monkeypatch):

332 """A producer-only scheduler must never populate meta.loads, even when

338 - entries are ready.

339 - """

333 + entries are ready."""

334 s = _make_scheduler(monkeypatch, ec_role="ec_producer")

335 req = _Request([_Feature("h1", length=1)])

336

@@ -378,8 +372,7 @@ def test_consumer_only_has_cache_item(monkeypatch):

372

373 def test_save_not_emitted_for_already_cached_entry(monkeypatch):

374 """An entry already in the cache (from a prior save) must not trigger

381 - a second allocation or appear in meta.saves again.

382 - """

375 + a second allocation or appear in meta.saves again."""

376 s = _make_scheduler(monkeypatch)

377 req = _Request([_Feature("h1", length=1)])

378

@@ -400,8 +393,7 @@ def test_save_not_emitted_for_already_cached_entry(monkeypatch):

393

394 def test_update_connector_output_ignores_foreign_meta(monkeypatch):

395 """A worker output whose payload is not an ECCPUWorkerMetadata is a

403 - no-op — no cache mutation, no crash.

404 - """

396 + no-op — no cache mutation, no crash."""

397 s = _make_scheduler(monkeypatch)

398 _seed_cached(s, "a", n_blocks=1)

399

@@ -415,8 +407,7 @@ def test_update_connector_output_ignores_foreign_meta(monkeypatch):

407

408 def test_update_connector_output_ignores_unknown_report(monkeypatch):

409 """Reports for a never-seen save hash or a transfer id this scheduler never

418 - dispatched are dropped rather than mutating unrelated state.

419 - """

410 + dispatched are dropped rather than mutating unrelated state."""

411 s = _make_scheduler(monkeypatch)

412 s.update_connector_output(_WorkerOutput(saves=["gone"], loads=[4242]))

413 assert s.has_cache_item("gone") is False

@@ -428,8 +419,7 @@ def test_update_connector_output_ignores_unknown_report(monkeypatch):

419

420 def test_has_pending_push_work_tracks_inflight_save(monkeypatch):

421 """A dispatched-but-unconfirmed save keeps push work pending until the

431 - completion report marks it ready.

432 - """

422 + completion report marks it ready."""

423 s = _make_scheduler(monkeypatch)

424 assert s.has_pending_push_work() is False

425

@@ -444,8 +434,7 @@ def test_has_pending_push_work_tracks_inflight_save(monkeypatch):

434

435 def test_has_pending_push_work_tracks_inflight_load(monkeypatch):

436 """A dispatched-but-unconfirmed load keeps push work pending until the

447 - unpin completion report arrives.

448 - """

437 + unpin completion report arrives."""

438 s = _make_scheduler(monkeypatch)

439 _seed_cached(s, "a", n_blocks=1)

440 assert s.has_pending_push_work() is False # ready + unpinned

@@ -464,8 +453,7 @@ def test_has_pending_push_work_tracks_inflight_load(monkeypatch):

453

454 def test_shutdown_disables_roles_and_cleans_region(monkeypatch):

455 """After shutdown the scheduler serves no items and the region is

467 - cleaned up.

468 - """

456 + cleaned up."""

457 s = _make_scheduler(monkeypatch, num_blocks=2)

458 _seed_cached(s, "a", n_blocks=2)

459 assert s.has_cache_item("a") is True

tests/v1/ec_connector/unit/cpu/test_ec_shared_region.py

+4 -8

@@ -98,8 +98,7 @@ def test_second_instance_opens_existing_file_and_shares_memory():

98 def test_only_creator_unlinks_file_on_cleanup():

99 """Critical contract: if the non-creator unlinks, the creator's mmap path

100 becomes a dangling backing file and a third opener would create a new one

101 - out from under the creator.

102 - """

101 + out from under the creator."""

102 instance_id = str(uuid.uuid4())

103 r1 = ECSharedRegion(

104 engine_id=instance_id,

@@ -216,8 +215,7 @@ def test_madvise_einval_selects_fallback_for_whole_region(monkeypatch):

215

216 def test_non_creator_does_not_probe_or_populate(monkeypatch):

217 """Only the creator pre-faults. A second opener must not probe the advice

219 - nor re-touch pages the creator already populated.

220 - """

218 + nor re-touch pages the creator already populated."""

219 from vllm.distributed.ec_transfer.ec_connector.cpu import ec_shared_region as esr

220

221 instance_id = str(uuid.uuid4())

@@ -314,8 +312,7 @@ def test_madvise_unexpected_oserror_propagates(monkeypatch):

312

313 def test_pin_memory_success_sets_flag(region):

314 """When cudaHostRegister returns 0, _is_pinned flips to True

317 - and cleanup will correspondingly call cudaHostUnregister.

318 - """

315 + and cleanup will correspondingly call cudaHostUnregister."""

316 fake_cudart = MagicMock()

317 success = MagicMock()

318 success.value = 0

@@ -335,8 +332,7 @@ def test_pin_memory_success_sets_flag(region):

332

333 def test_pin_memory_failure_leaves_flag_false():

334 """If cudaHostRegister fails (non-zero), don't pretend it succeeded —

338 - cleanup must NOT call cudaHostUnregister on memory we never registered.

339 - """

335 + cleanup must NOT call cudaHostUnregister on memory we never registered."""

336 r = _make_region()

337 try:

338 fake_cudart = MagicMock()

tests/v1/ec_connector/unit/cpu/worker/test_worker.py

+6 -12

@@ -363,8 +363,7 @@ def test_save_caches_coalesces_only_adjacent_blocks(

363

364 def test_save_caches_noop_when_mm_hash_not_in_saves(make_worker):

365 """When the scheduler hasn't pre-allocated blocks for ``mm_hash``,

366 - ``save_caches`` + ``flush_saves`` is a pure no-op.

367 - """

366 + ``save_caches`` + ``flush_saves`` is a pure no-op."""

367 worker = make_worker()

368 sentinel = 0x42

369 worker._region.blocks.fill_(sentinel)

@@ -382,8 +381,7 @@ def test_save_caches_noop_when_mm_hash_not_in_saves(make_worker):

381 )

382 def test_save_caches_noop_for_non_save_rank(make_worker, tp_rank, pcp_rank):

383 """Only TP rank 0 + PCP rank 0 writes to mmap. All other combinations

385 - must skip the save entirely.

386 - """

384 + must skip the save entirely."""

385 worker = make_worker(tp_rank=tp_rank, pcp_rank=pcp_rank)

386 sentinel = 0x42

387 worker._region.blocks.fill_(sentinel)

@@ -398,8 +396,7 @@ def test_save_caches_noop_for_non_save_rank(make_worker, tp_rank, pcp_rank):

396

397 def test_save_caches_raises_when_allocated_blocks_too_small(make_worker):

398 """``save_caches`` must raise ``AssertionError`` when the encoder output

401 - is larger than the allocated block space.

402 - """

399 + is larger than the allocated block space."""

400 worker = make_worker()

401 src = torch.zeros(3 * _HIDDEN_DIM, dtype=_DTYPE)

402 with pytest.raises(AssertionError, match="exceeds allocated blocks"):

@@ -552,8 +549,7 @@ def test_start_load_caches_noop_when_loads_is_empty(make_worker):

549 def test_start_load_caches_skips_cached_and_loads_new_in_same_step(make_worker):

550 """Every hash in ``meta.loads`` is copied from mmap, including one whose

551 key is already resident in ``encoder_cache`` — there is no worker-side

555 - skip.

556 - """

552 + skip."""

553 worker = make_worker()

554 n_blocks = 3

555 src_orig = torch.arange(n_blocks * _HIDDEN_DIM, dtype=_DTYPE).reshape(

@@ -782,8 +778,7 @@ def test_buffer_pool_is_reused_across_save_steps(make_worker):

778 @_requires_accelerator

779 def test_buffer_pool_is_reused_across_load_steps(make_worker):

780 """Once a load copy completes its descriptor buffers return to the pool and

785 - are reused by the next call — no reallocation.

786 - """

781 + are reused by the next call — no reallocation."""

782 worker = make_worker()

783 worker._region.blocks[0].fill_(0x01)

784 worker._region.blocks[1].fill_(0x02)

@@ -854,8 +849,7 @@ def test_init_pins_memory_when_available(pin_available, expected_pinned):

849

850 def test_shutdown_calls_region_cleanup_and_swallows_errors(caplog_vllm):

851 """``shutdown`` must always call ``region.cleanup`` — and must never

857 - raise.

858 - """

852 + raise."""

853 worker = object.__new__(ECCPUWorker)

854 mock_region = Mock(spec=ECSharedRegion)

855 worker._region = mock_region

tests/v1/ec_connector/unit/test_epd_proxy_round_robin.py

+1 -2

@@ -43,8 +43,7 @@ def assign(proxy):

43

44 def _drive(assign, e_urls, counts):

45 """Feed a sequence of request item-counts through the cursor, as

46 - fanout_encoder_primer does one request at a time.

47 - """

46 + fanout_encoder_primer does one request at a time."""

47 cursor = 0

48 all_urls = []

49 for count in counts:

tests/v1/ec_connector/unit/test_metadata.py

+1 -2

@@ -40,8 +40,7 @@ def test_metadata_saves_dict_operations():

40

41 def test_metadata_loads_carry_a_transfer_id_and_blocks():

42 """Each load entry is a (transfer_id, block_ids) pair, so the worker can

43 - report the transfer rather than the mm_hash.

44 - """

43 + report the transfer rather than the mm_hash."""

44 meta = ECCPUConnectorMetadata()

45

46 meta.loads["mm_hash_1"] = (0, [5, 6, 7])

tests/v1/ec_connector/unit/test_scheduler_nixl_producer.py

+1 -2

@@ -79,8 +79,7 @@ def test_request_finished_announces_not_ready_entry(monkeypatch):

79 """A save whose GPU->mmap copy hasn't been confirmed complete yet is

80 still announced: the entry can't be evicted before it's ready, and a read

81 arriving too early is NACKed NACK_NOT_READY for the consumer to retry —

82 - protecting it is not this method's job.

83 - """

82 + protecting it is not this method's job."""

83 s = _sched_gate_off(monkeypatch)

84 s._nixl_enabled = True

85 s._peer_host, s._peer_port = "1.2.3.4", 5601

tests/v1/ec_connector/unit/test_session.py

+2 -4

@@ -129,8 +129,7 @@ def test_producer_session_grant_nacks_missing_entry():

129

130 def test_producer_session_grant_nacks_not_ready_entry_as_retryable():

131 """An announced entry whose save is still in flight is distinguished from a

132 - miss, so the consumer retries instead of falling back to local compute.

133 - """

132 + miss, so the consumer retries instead of falling back to local compute."""

133 cache = EmbeddingCache(num_blocks=8)

134 cache.alloc("h1", 2) # not ready

135 s = _make_producer_session(cache)

@@ -438,8 +437,7 @@ def test_consumer_session_poll_nack_goes_to_tombstoned():

437

438 def test_consumer_session_poll_not_ready_nack_goes_to_retryable():

439 """A not-ready NACK must not tombstone: tombstoning admits the request with

441 - no embedding, which is fatal once the media has been rewritten away.

442 - """

440 + no embedding, which is fatal once the media has been rewritten away."""

441 import msgspec

442

443 s = _make_consumer_session()

tests/v1/engine/test_core_engine_actor_manager.py

+1 -2

@@ -68,8 +68,7 @@ class _StubEngineCoreActor(EngineCoreActorMixin):

68 # in ``tests/entrypoints/test_api_server_process_manager.py``.

69 def _bind_and_report_worker(listen_address, sock, args, client_config):

70 """Bind ROUTER/PULL with a kernel-assigned port, report the actual

71 - endpoints back via ``actual_address_pipe``, then exit.

72 - """

71 + endpoints back via ``actual_address_pipe``, then exit."""

72 ctx = zmq.Context()

73 try:

74 in_sock = make_zmq_socket(

tests/v1/engine/test_engine_core.py

+4 -8

@@ -607,8 +607,7 @@ def _cadenced_dp_engine_core(

607 monkeypatch, results: list[tuple[bool, bool]], dp_sync_interval: int = 32

608 ):

609 """A bare DPEngineCoreProc whose all-reduce is scripted by `results`;

610 - returns the core and the step numbers at which the all-reduce ran.

611 - """

610 + returns the core and the step numbers at which the all-reduce ran."""

611 core = object.__new__(DPEngineCoreProc)

612 core.dp_group = object()

613 core.dp_sync_interval = dp_sync_interval

@@ -627,8 +626,7 @@ def _cadenced_dp_engine_core(

626

627 def test_dp_sync_interval_default_is_16():

628 """Regression pin: lowering the default narrows the mid-wave pause tail

630 - for async RL, where the engine is rarely idle when pause lands.

631 - """

629 + for async RL, where the engine is rarely idle when pause lands."""

630 assert ParallelConfig.dp_sync_interval == 16

631

632

@@ -646,8 +644,7 @@ def test_dp_sync_interval_normal_wave(monkeypatch, dp_sync_interval: int):

644 def test_dp_sync_interval_idle_pause_consensus_on_first_step(monkeypatch):

645 """A pause of an idle engine arms every rank before its kick-started

646 first step, so the step-1 sync reaches consensus after one dummy batch

649 - regardless of the configured cadence.

650 - """

647 + regardless of the configured cadence."""

648 core, synced = _cadenced_dp_engine_core(

649 monkeypatch, [(False, True)], dp_sync_interval=32

650 )

@@ -673,8 +670,7 @@ def _pausable_engine_core_proc() -> EngineCoreProc:

670 @pytest.mark.parametrize("deferred", [False, True])

671 def test_pause_synchronizes_device_before_cache_reset(deferred: bool):

672 """A resolved pause promises an idle device: the barrier must run before

676 - caches are cleared and before the caller is unblocked.

677 - """

673 + caches are cleared and before the caller is unblocked."""

674 core = _pausable_engine_core_proc()

675 core.engines_running = deferred

676 order: list[str] = []

tests/v1/engine/test_engine_core_client.py

+3 -6

@@ -249,8 +249,7 @@ def test_dplb_non_late_interaction_still_uses_lb():

249

250 def test_dplb_burst_round_robins_despite_snapshot_rebinds():

251 """A stats snapshot rebind wipes the optimistic lb_engines increments;

252 - the exact in-flight floor must keep a burst spreading round-robin.

253 - """

252 + the exact in-flight floor must keep a burst spreading round-robin."""

253 client = _make_dplb_client(num_engines=4)

254

255 for _ in range(4):

@@ -325,8 +324,7 @@ async def test_dplb_scale_down_routes_after_stale_stats_snapshot():

324

325 def test_dplb_snapshot_backpressure_overrides_inflight():

326 """An engine reported heavily loaded by the coordinator is avoided even

328 - when this client has routed nothing to it.

329 - """

327 + when this client has routed nothing to it."""

328 client = _make_dplb_client(num_engines=2)

329 client.lb_engines = [[5, 10, 0.0], [0, 0, 0.0]]

330

@@ -339,8 +337,7 @@ def test_dplb_snapshot_backpressure_overrides_inflight():

337

338 def test_dplb_kv_pressure_amplifies_waiting_penalty():

339 """A waiting queue on a KV-bound engine (slow drain) is penalized, while

342 - the same queue with low KV usage is not (e.g. transient burst).

343 - """

340 + the same queue with low KV usage is not (e.g. transient burst)."""

341 client = _make_dplb_client(num_engines=2)

342 # Engine 0 has a smaller total but is KV-bound with a queue.

343 client.lb_engines = [[5, 10, 1.0], [0, 20, 0.2]]

tests/v1/engine/test_logprobs_processor.py

+1 -2

@@ -30,8 +30,7 @@ def _make_processor(num_logprobs: int) -> LogprobsProcessor:

30

31 def test_drops_trailing_sentinel_columns():

32 """A request that asked for 3 custom token logprobs but ended up in a

33 - batch padded to width 5 must not surface the trailing -inf entries.

34 - """

33 + batch padded to width 5 must not surface the trailing -inf entries."""

34 processor = _make_processor(num_logprobs=3)

35

36 sampled = 42

tests/v1/engine/test_output_processor.py

+1 -2

@@ -229,8 +229,7 @@ def test_request_stream_interval_raises_but_not_below_engine_default(

229 ):

230 """A per-request stream_interval can raise the interval above the engine

231 default but not below it (values under the default clamp up), without

232 - altering the generated text.

233 - """

232 + altering the generated text."""

233 engine_stream_interval = 5

234 # Request 0 (below the default) clamps up to 5; request 1 raises it to 10.

235 request_stream_intervals = [1, 10]

tests/v1/kv_connector/unit/offloading_connector/test_canonical_mapping.py

+1 -2

@@ -80,8 +80,7 @@ def _split_hnd_cache(spec) -> torch.Tensor:

80

81 def _packed_nhd_cache(spec) -> torch.Tensor:

82 """Logical (num_blocks, heads, block_size, 2 * head_size) over an NHD

83 - physical layout — the FlashAttention/FlashInfer/Triton/Flex form.

84 - """

83 + physical layout — the FlashAttention/FlashInfer/Triton/Flex form."""

84 return torch.zeros(

85 NUM_BLOCKS,

86 spec.block_size,

tests/v1/kv_connector/unit/offloading_connector/test_config.py

+2 -4

@@ -863,8 +863,7 @@ def test_parallelism_agnostic_excluded(kv_cache_groups: list[KVCacheGroupSpec]):

863 )

864 def test_canonical_layout_gate(kv_cache_groups, certified):

865 """The canonical layout certifies portability group by group; none of

866 - these shapes are portable in the direct layout.

867 - """

866 + these shapes are portable in the direct layout."""

867 assert not _parallelism_agnostic(kv_cache_groups)

868 assert _parallelism_agnostic(kv_cache_groups, canonical=True) is certified

869

@@ -872,8 +871,7 @@ def test_canonical_layout_gate(kv_cache_groups, certified):

871 def test_canonical_layout_certifies_v2_model_runner():

872 """Canonical bytes are certified per layer against live tensor strides at

873 registration, so the static gate must not depend on the model-runner

875 - version — the v2 runner is the case the canonical layout exists for.

876 - """

874 + version — the v2 runner is the case the canonical layout exists for."""

875 groups = _groups(_full_attention_spec())

876 assert _parallelism_agnostic(groups, canonical=True, v2=True)

877

tests/v1/kv_connector/unit/offloading_connector/test_metrics.py

+2 -4

@@ -117,8 +117,7 @@ def _spec_cls_with_metric_definitions(

117 metric_definitions: dict[str, Any],

118 ) -> type:

119 """Build a fake offloading spec class reporting the given metric

120 - definitions, so tests don't need to patch the real CPU spec.

121 - """

120 + definitions, so tests don't need to patch the real CPU spec."""

121

122 class _FakeOffloadingSpec:

123 @staticmethod

@@ -188,8 +187,7 @@ def test_build_kv_connector_stats_with_empty_dict():

187

188 def test_build_kv_connector_stats_reconstructs_offload_stats():

189 """Test that OffloadingConnector stats are properly reconstructed with

191 - correct data.

192 - """

190 + correct data."""

191 serialized_data = {

192 _StatsKey.TYPES: {

193 LOAD_BYTES: _MetricType.COUNTER,

tests/v1/kv_connector/unit/offloading_connector/test_scheduler.py

+16 -32

@@ -1583,8 +1583,7 @@ _SCAN_BEHAVIOR = {

1583 @pytest.mark.parametrize("result", list(LookupResult))

1584 def test_scan_behavior_declared_for_every_lookup_result(result: LookupResult):

1585 """Whether a result keeps a sliding-window streak alive decides how wide

1586 - the demanded chunk set gets, so every member needs deliberate behavior.

1587 - """

1586 + the demanded chunk set gets, so every member needs deliberate behavior."""

1587 assert result in _SCAN_BEHAVIOR, f"{result} has no declared scan behavior"

1588 expected_lookups, expected_end = _SCAN_BEHAVIOR[result]

1589

@@ -2035,8 +2034,7 @@ def test_request_level_policy_stores_all_blocks(request_runner, async_scheduling

2034

2035 def test_loads_do_not_populate_fence_index(request_runner):

2036 """Loads don't populate _block_id_to_pending_jobs (protected by

2038 - delay_free_blocks while in flight).

2039 - """

2037 + delay_free_blocks while in flight)."""

2038 runner = request_runner(

2039 blocks_per_chunk=3,

2040 block_size=4,

@@ -2106,8 +2104,7 @@ def test_fence_at_update_state_after_alloc(request_runner):

2104 def test_fence_at_build_store_jobs(request_runner):

2105 """A new prefill (no load -> update_state_after_alloc returns early)

2106 reusing a finished request's pending-store block is flushed by

2109 - _build_store_jobs's fence.

2110 - """

2107 + _build_store_jobs's fence."""

2108 runner = request_runner(

2109 blocks_per_chunk=1,

2110 block_size=4,

@@ -2157,8 +2154,7 @@ def test_fence_at_build_store_jobs(request_runner):

2154 @pytest.mark.parametrize("async_scheduling", [True, False])

2155 def test_complete_store_called_per_job(request_runner, async_scheduling: bool):

2156 """complete_store fires per-job, not deferred to request finish.

2160 - Each call carries only that store's keys.

2161 - """

2157 + Each call carries only that store's keys."""

2158 tokens_per_block = 4

2159 blocks_per_chunk = 3

2160 tokens_per_chunk = tokens_per_block * blocks_per_chunk

@@ -2387,8 +2383,7 @@ def test_offload_prompt_only(request_runner, async_scheduling: bool):

2383 @pytest.mark.parametrize("async_scheduling", [True, False])

2384 def test_reset_cache(request_runner, async_scheduling: bool):

2385 """reset_cache flushes in-flight loads, calls manager.reset_cache(), resets

2390 - next_stored_chunk_idx for active requests and clears job tracking.

2391 - """

2386 + next_stored_chunk_idx for active requests and clears job tracking."""

2387 block_size = 4

2388 blocks_per_chunk = 3

2389 tokens_per_chunk = block_size * blocks_per_chunk

@@ -2934,8 +2929,7 @@ def test_stale_sliding_window_block_after_prepare_store_failure(

2929 @pytest.mark.parametrize("async_scheduling", [True, False])

2930 def test_skip_reading_prefix_cache(request_runner, async_scheduling: bool):

2931 """When skip_reading_prefix_cache=True, the offloading connector must not

2937 - load any blocks from CPU even if a matching prefix is cached there.

2938 - """

2932 + load any blocks from CPU even if a matching prefix is cached there."""

2933 block_size = 4

2934 blocks_per_chunk = 3

2935 tokens_per_chunk = block_size * blocks_per_chunk

@@ -3029,8 +3023,7 @@ def test_max_load_tokens_limits_external_load(request_runner, async_scheduling:

3023

3024 class TestEagle:

3025 """Tests for Eagle/MTP speculative decoding support in the offloading

3032 - connector scheduler — both _lookup() unit tests and integration tests.

3033 - """

3026 + connector scheduler — both _lookup() unit tests and integration tests."""

3027

3028 # -------------------------------------------------------------------

3029 # Helpers

@@ -3572,8 +3565,7 @@ class TestEagle:

3565 self, request_runner, async_scheduling: bool

3566 ):

3567 """Eagle sliding-window group stores all prompt blocks but excludes

3575 - the trailing chunk while decoding.

3576 - """

3568 + the trailing chunk while decoding."""

3569 block_size = 4

3570 sliding_window = 8

3571 num_gpu_blocks = 100

@@ -3625,8 +3617,7 @@ class TestEagle:

3617 self, request_runner, async_scheduling: bool

3618 ):

3619 """An eagle group with a single-block prompt stores it at the end of

3628 - prefill: prompt blocks are stable, so no tail is held back.

3629 - """

3620 + prefill: prompt blocks are stable, so no tail is held back."""

3621 block_size = 4

3622 blocks_per_chunk = 1

3623 tokens_per_chunk = block_size * blocks_per_chunk

@@ -4263,8 +4254,7 @@ def test_request_finished_mixed_full_attn_and_sliding_window(

4254

4255 def test_chunked_local_attention_reports_its_chunk_window():

4256 """Llama 4 uses chunked local attention, which used to trip the

4266 - FullAttentionSpec assert and kill the engine at startup.

4267 - """

4257 + FullAttentionSpec assert and kill the engine at startup."""

4258 spec = ChunkedLocalAttentionSpec(

4259 block_size=16,

4260 num_kv_heads=2,

@@ -4503,8 +4493,7 @@ def test_retention_interval_zero_stores_only_replay_boundary(

4493 def _shared_kv_mtp_config():

4494 """Speculative config for a shared-group MTP model: eagle-family method

4495 whose drafter layer merges into a target KV-cache group, so no group

4506 - self-identifies as a drafter group (issue #52735).

4507 - """

4496 + self-identifies as a drafter group (issue #52735)."""

4497 spec = MagicMock(name="shared_kv_mtp_spec")

4498 spec.use_eagle.return_value = True

4499 spec.use_eagle_block_drop.return_value = True

@@ -4518,14 +4507,12 @@ def _shared_kv_mtp_config():

4507 class TestSharedGroupMTPOffload:

4508 """Regression tests for issue #52735: OffloadingConnector must keep

4509 serving when speculative decoding is enabled but no KV-cache group is

4521 - annotated as a drafter group (shared-group MTP models).

4522 - """

4510 + annotated as a drafter group (shared-group MTP models)."""

4511

4512 def test_no_annotation_marks_no_groups(self, request_runner):

4513 """Spec decode on + zero annotated groups must NOT mark every group

4514 as a drafter group; the full store->load roundtrip must match the

4527 - non-speculative behavior of test_two_groups_full_and_sliding_window.

4528 - """

4515 + non-speculative behavior of test_two_groups_full_and_sliding_window."""

4516 block_size = 4

4517 kv_cache_groups = [

4518 KVCacheGroupSpec(

@@ -4582,8 +4569,7 @@ class TestSharedGroupMTPOffload:

4569 class TestMambaHybridOffloadServing:

4570 """Store->finish->lookup flows on a full-attention + mamba-align hybrid,

4571 with a manager that only HITs keys that were actually stored. Guards the

4585 - two collapse routes of issue #52735.

4586 - """

4572 + two collapse routes of issue #52735."""

4573

4574 BLOCK = 4

4575 MAMBA_BLOCK = 16

@@ -4712,8 +4698,7 @@ class TestMambaHybridOffloadServing:

4698 """Route 1 of #52735: with spec decode on and no drafter annotation,

4699 serving must equal the non-speculative baseline (was 0 before the

4700 fix: the all-groups fallback marked the mamba group as a drafter and

4715 - the volatile-tail pop consumed its only servable chunk).

4716 - """

4701 + the volatile-tail pop consumed its only servable chunk)."""

4702 baseline = self._roundtrip_served_tokens(self._make_scheduler(None))

4703 assert baseline == 16

4704 served = self._roundtrip_served_tokens(

@@ -4726,8 +4711,7 @@ class TestMambaHybridOffloadServing:

4711 group must not drag the confirmed boundary below a coarser sibling's

4712 chunk granularity. The widened query makes the volatile-tail pop

4713 land on an extra queried chunk, holding the boundary at 16 tokens

4729 - (was 0 before the fix: 4 hits -> pop -> 12 tokens < mamba chunk).

4730 - """

4714 + (was 0 before the fix: 4 hits -> pop -> 12 tokens < mamba chunk)."""

4715 scheduler = self._make_scheduler(None, mamba_eagle=True)

4716 assert [c.is_eagle_group for c in scheduler.config.kv_group_configs] == [

4717 True,

tests/v1/kv_connector/unit/offloading_connector/test_worker.py

+1 -2

@@ -80,8 +80,7 @@ def _allocate_kv_caches(

80

81 def _single_rank_vllm_config(total_kv_heads: int):

82 """A one-rank (TP=1) parallel config, as canonical mappings are derived

83 - from it.

84 - """

83 + from it."""

84 vllm_config = MagicMock()

85 parallel_config = vllm_config.parallel_config

86 parallel_config.tensor_parallel_size = 1

tests/v1/kv_connector/unit/test_bidirectional_kv_transfer.py

+19 -38

@@ -148,8 +148,7 @@ def test_multiturn_lifecycle():

148 (do_remote_decode=True). Finishes LENGTH_CAPPED with remote_block_ids.

149 Turn 2: P receives remote_block_ids from D. P pulls KV from D because

150 remote_block_ids is not None and external tokens > 0. Computes only

151 - new tokens, finishes LENGTH_CAPPED.

152 - """

151 + new tokens, finishes LENGTH_CAPPED."""

152 vllm_config = create_vllm_config(

153 kv_connector_extra_config=BIDIR_KV_EXTRA_CONFIG,

154 )

@@ -195,8 +194,7 @@ def test_multiturn_lifecycle():

194

195 def test_first_turn_no_remote_blocks():

196 """First turn: P has no remote_block_ids from D yet.

198 - Standard local prefill, returns kv_transfer_params for future turns.

199 - """

197 + Standard local prefill, returns kv_transfer_params for future turns."""

198 vllm_config = create_vllm_config(

199 kv_connector_extra_config=BIDIR_KV_EXTRA_CONFIG,

200 )

@@ -276,8 +274,7 @@ def test_abort_p_side_non_length_capped():

274

275 def test_remote_blocks_exceed_prompt_tokens():

276 """D provides more remote tokens than P's prompt needs.

279 - P caps external tokens to prompt length.

280 - """

277 + P caps external tokens to prompt length."""

278 vllm_config = create_vllm_config(

279 kv_connector_extra_config=BIDIR_KV_EXTRA_CONFIG,

280 )

@@ -313,8 +310,7 @@ def test_remote_blocks_exceed_prompt_tokens():

310 def test_p_node_pulls_partial_last_block_from_d():

311 """D sends remote_block_ids with partially filled last block.

312 remote_num_tokens < len(remote_block_ids) * block_size.

316 - P pulls only remote_num_tokens worth of external tokens.

317 - """

313 + P pulls only remote_num_tokens worth of external tokens."""

314 vllm_config = create_vllm_config(

315 kv_connector_extra_config=BIDIR_KV_EXTRA_CONFIG,

316 )

@@ -359,8 +355,7 @@ def test_p_node_pulls_partial_last_block_from_d():

355

356 def test_add_new_req_to_recv_populates_remote_meta():

357 """add_new_req_to_recv correctly populates RemoteMeta for P-node

362 - bi-directional KV pull from D.

363 - """

358 + bi-directional KV pull from D."""

359 meta = NixlConnectorMetadata()

360 kv_params = {

361 "remote_block_ids": [[0, 1, 2]],

@@ -388,8 +383,7 @@ def test_add_new_req_to_recv_populates_remote_meta():

383

384 def test_build_connector_meta_recv_entries():

385 """P-node scheduler: do_remote_decode=True + remote_block_ids →

391 - _reqs_need_recv populated, build_connector_meta produces reqs_to_recv.

392 - """

386 + _reqs_need_recv populated, build_connector_meta produces reqs_to_recv."""

387 vllm_config = create_vllm_config(

388 kv_connector_extra_config=BIDIR_KV_EXTRA_CONFIG,

389 )

@@ -463,8 +457,7 @@ def test_p_node_pull_kv_from_d(dist_init):

457 )

458 def test_p_node_pull_then_send_kv(dist_init):

459 """Full P-node bi-directional: pull KV from D → prefill →

466 - send KV back to D via notification.

467 - """

460 + send KV back to D via notification."""

461 connector, worker = _make_connector_with_fake_worker()

462 meta = _make_p_node_recv_metadata("req-p2", [10, 11], [20, 21])

463 _do_load_kv(connector, meta)

@@ -511,8 +504,7 @@ def test_p_node_deferred_pull_on_no_handshake(dist_init):

504 def test_d_node_request_finished_returns_kv_params():

505 """D-node request_finished returns kv_transfer_params with

506 do_remote_decode=True, remote_block_ids, remote_num_tokens

514 - for P to pull. These params go directly to P node.

515 - """

507 + for P to pull. These params go directly to P node."""

508 vllm_config = create_vllm_config(

509 kv_connector_extra_config=BIDIR_KV_EXTRA_CONFIG,

510 )

@@ -593,8 +585,7 @@ def test_d_node_request_finished_remote_num_tokens():

585

586 def test_d_node_partial_last_block_remote_num_tokens():

587 """D-node: remote_num_tokens < len(remote_block_ids) * block_size

596 - when last block is partially filled.

597 - """

588 + when last block is partially filled."""

589 vllm_config = create_vllm_config(

590 kv_connector_extra_config=BIDIR_KV_EXTRA_CONFIG,

591 )

@@ -628,8 +619,7 @@ def test_no_double_read_blocks_after_reschedule():

619 bidirectional request (once on initial schedule, once after

620 WAITING_FOR_REMOTE_KVS → reschedule). The _remote_blocks_processed

621 flag must prevent the request from being added to _reqs_need_recv

631 - twice, which would cause P to read D's blocks twice.

632 - """

622 + twice, which would cause P to read D's blocks twice."""

623 vllm_config = create_vllm_config(

624 kv_connector_extra_config=BIDIR_KV_EXTRA_CONFIG,

625 )

@@ -682,8 +672,7 @@ def test_remote_num_tokens_bounded_by_blocks():

672 """Edge case 2: D-node request_finished must return

673 remote_num_tokens <= len(remote_block_ids) * block_size.

674 request.num_tokens includes the last sampled token which has no KV

685 - in the cache, so remote_num_tokens must use num_computed_tokens.

686 - """

675 + in the cache, so remote_num_tokens must use num_computed_tokens."""

676 vllm_config = create_vllm_config(

677 kv_connector_extra_config=BIDIR_KV_EXTRA_CONFIG,

678 )

@@ -719,8 +708,7 @@ def test_remote_num_tokens_bounded_by_blocks():

708 def test_kv_recompute_threshold_skips_small_transfer():

709 """Edge case 3: When remote tokens are below kv_recompute_threshold,

710 P should skip the remote pull and compute locally instead of

722 - entering WAITING_FOR_REMOTE_KVS.

723 - """

711 + entering WAITING_FOR_REMOTE_KVS."""

712 threshold = 256

713 vllm_config = create_vllm_config(

714 kv_connector_extra_config={

@@ -762,8 +750,7 @@ def test_p_node_finished_holds_blocks_for_d():

750 """Edge case 4: P-node finishes with FINISHED_LENGTH_CAPPED and

751 do_remote_decode=True. P must hold blocks (delay_free=True) and

752 return kv_transfer_params with do_remote_prefill=True so D can

765 - read P's blocks.

766 - """

753 + read P's blocks."""

754 vllm_config = create_vllm_config(

755 kv_connector_extra_config=BIDIR_KV_EXTRA_CONFIG,

756 )

@@ -804,8 +791,7 @@ def test_p_node_finished_holds_blocks_for_d():

791 def test_cache_miss_first_turn_no_remote_pull():

792 """Edge case 5: First turn with do_remote_decode=True but no

793 remote_block_ids (cache MISS). P should prefill locally with

807 - num_external_tokens=0 and not enter WAITING_FOR_REMOTE_KVS.

808 - """

794 + num_external_tokens=0 and not enter WAITING_FOR_REMOTE_KVS."""

795 vllm_config = create_vllm_config(

796 kv_connector_extra_config=BIDIR_KV_EXTRA_CONFIG,

797 )

@@ -840,8 +826,7 @@ def test_cache_miss_first_turn_no_remote_pull():

826 def test_partial_remote_tokens_less_than_prompt():

827 """Edge case 6: D's remote_num_tokens covers only part of P's

828 prompt. P should pull remote_num_tokens worth of external tokens

843 - and compute the rest locally.

844 - """

829 + and compute the rest locally."""

830 vllm_config = create_vllm_config(

831 kv_connector_extra_config=BIDIR_KV_EXTRA_CONFIG,

832 )

@@ -886,8 +871,7 @@ def test_partial_remote_tokens_less_than_prompt():

871 def test_remote_blocks_processed_flag_persists():

872 """Edge case 7: After recv completes and request is rescheduled,

873 the _remote_blocks_processed flag in kv_transfer_params prevents

889 - the bidirectional path from re-entering _reqs_need_recv.

890 - """

874 + the bidirectional path from re-entering _reqs_need_recv."""

875 vllm_config = create_vllm_config(

876 kv_connector_extra_config=BIDIR_KV_EXTRA_CONFIG,

877 )

@@ -959,8 +943,7 @@ _REMOTE = FakeNixlConnectorWorker.REMOTE_ENGINE_ID

943 )

944 def test_turn2_deadline_gate(dist_init, offset, expiry_delta, expect_declined):

945 """P declines on the turn-2 readback if the deadline is known,

962 - near-expiry and a handshake clock offset is known

963 - """

946 + near-expiry and a handshake clock offset is known"""

947 connector, worker = _make_connector_with_fake_worker()

948 worker._engine_clock_offset[_REMOTE] = offset

949 expiry_time = None if expiry_delta is None else time.perf_counter() + expiry_delta

@@ -996,8 +979,7 @@ def test_turn2_deadline_gate(dist_init, offset, expiry_delta, expect_declined):

979 )

980 def test_turn2_full_prefix_hit_with_expired_deadline_skips_gate(dist_init):

981 """A full local prefix hit issues no READ. The expiry gate must be skipped

999 - even when D's deadline is expired.

1000 - """

982 + even when D's deadline is expired."""

983 connector, worker = _make_connector_with_fake_worker()

984 worker._engine_clock_offset[_REMOTE] = 0.0

985

@@ -1050,8 +1032,7 @@ def test_d_node_request_finished_exports_blocks_expiry_time():

1032

1033 def test_handshake_listener_appends_perf_counter_frame():

1034 """The handshake reply carries a 2nd frame with the listener's live

1053 - perf_counter.

1054 - """

1035 + perf_counter."""

1036 import threading

1037

1038 import msgspec

tests/v1/kv_connector/unit/test_config.py

+6 -12

@@ -19,15 +19,13 @@ class _StubLMCacheMPConnector:

19 test only asserts on the connector *name* and the ``extra_config`` dict

20 produced by ``VllmConfig``, never instantiates the connector, so a bare

21 placeholder class is sufficient. Not subclassing ``SupportsHMA`` mirrors

22 - the real connector's HMA support (it does not support HMA either).

23 - """

22 + the real connector's HMA support (it does not support HMA either)."""

23

24

25 @pytest.fixture

26 def stub_lmcache_mp_connector(monkeypatch):

27 """Replace the lazy loader so VllmConfig.__post_init__ does not import

29 - ``lmcache_mp_connector`` (and thus ``lmcache``) during config tests.

30 - """

28 + ``lmcache_mp_connector`` (and thus ``lmcache``) during config tests."""

29 monkeypatch.setitem(

30 KVConnectorFactory._registry,

31 "LMCacheMPConnector",

@@ -107,8 +105,7 @@ def _build_config(

105 enable_cumem_allocator: bool = False,

106 ) -> VllmConfig:

107 """Build a VllmConfig that exercises _verify_kv_transfer_compat without

110 - requiring a real model (avoids HF downloads in CI).

111 - """

108 + requiring a real model (avoids HF downloads in CI)."""

109 from types import SimpleNamespace

110

111 kv_transfer_config = (

@@ -134,8 +131,7 @@ def test_kv_connector_rejects_expandable_segments(monkeypatch, kv_connector):

131 invalidated when expandable_segments lets the CUDA VMM allocator remap

132 the underlying physical pages. We can't enumerate every connector that

133 does this (especially OOT ones), so reject the combination whenever any

137 - connector is configured.

138 - """

134 + connector is configured."""

135 monkeypatch.setenv("PYTORCH_CUDA_ALLOC_CONF", "expandable_segments:True")

136 with pytest.raises(ValueError, match="expandable_segments"):

137 _build_config(kv_connector=kv_connector)

@@ -143,8 +139,7 @@ def test_kv_connector_rejects_expandable_segments(monkeypatch, kv_connector):

139

140 def test_kv_connector_allows_expandable_segments_with_sleep_mode(monkeypatch):

141 """Sleep mode routes KV allocations through CuMemAllocator's pool, which

146 - auto-disables expandable_segments (see #40812).

147 - """

142 + auto-disables expandable_segments (see #40812)."""

143 monkeypatch.setenv("PYTORCH_CUDA_ALLOC_CONF", "expandable_segments:True")

144 _build_config(kv_connector="NixlConnector", enable_sleep_mode=True)

145

@@ -167,8 +162,7 @@ def test_kv_connector_allows_other_alloc_conf(monkeypatch):

162

163 def test_no_kv_connector_ignores_expandable_segments(monkeypatch):

164 """The expandable_segments check only applies when a KV connector is

170 - configured.

171 - """

165 + configured."""

166 monkeypatch.setenv("PYTORCH_CUDA_ALLOC_CONF", "expandable_segments:True")

167 _build_config(kv_connector=None)

168

tests/v1/kv_connector/unit/test_flexkv_connector.py

+3 -6

@@ -50,8 +50,7 @@ def _make_flexkv_module(

50 impl_mock: MagicMock,

51 ) -> tuple[types.ModuleType, types.ModuleType]:

52 """Build a fake ``flexkv`` package hierarchy that returns *impl_mock*

53 - when ``FlexKVConnectorV1Impl`` is instantiated.

54 - """

53 + when ``FlexKVConnectorV1Impl`` is instantiated."""

54 flexkv_mod = types.ModuleType("flexkv")

55 integration_mod = types.ModuleType("flexkv.integration")

56 vllm_mod = types.ModuleType("flexkv.integration.vllm")

@@ -72,8 +71,7 @@ def _make_flexkv_module(

71

72 def _install_flexkv_mock(impl_mock: MagicMock):

73 """Insert fake flexkv modules into sys.modules and return a context that

75 - cleans them up afterwards.

76 - """

74 + cleans them up afterwards."""

75 flexkv_mod, adapter_mod = _make_flexkv_module(impl_mock)

76 mods = {

77 "flexkv": flexkv_mod,

@@ -106,8 +104,7 @@ def _build_connector(vllm_config: VllmConfig, impl_mock: MagicMock):

104

105 class TestFlexKVConnectorImportError:

106 """FlexKVConnectorV1 should fail with a helpful message when flexkv is

109 - absent.

110 - """

107 + absent."""

108

109 def test_import_error_message(self):

110 from vllm.distributed.kv_transfer.kv_connector.v1.flexkv_connector import (

tests/v1/kv_connector/unit/test_handshake_pp_aggregation.py

+3 -6

@@ -176,8 +176,7 @@ def test_engine_unwraps_handshake_metadata_for_legacy_connector(

176 ) -> None:

177 """Engine core always asks workers for `(pp_rank, tp_rank)`-keyed metadata,

178 then unwraps to `{tp_rank: metadata}` for a connector that has not opted

179 - into PP-aware handshake (single-PP producer, all `pp_rank == 0`).

180 - """

179 + into PP-aware handshake (single-PP producer, all `pp_rank == 0`)."""

180 metadata_0 = _Metadata()

181 metadata_1 = _Metadata()

182 connector = _LegacyConnector()

@@ -200,8 +199,7 @@ def test_engine_rejects_pp_producer_for_legacy_connector(

199 monkeypatch: pytest.MonkeyPatch,

200 ) -> None:

201 """A connector that has not opted into PP-aware handshake must not silently

203 - drop metadata from `pp_rank > 0`; engine core init raises instead.

204 - """

202 + drop metadata from `pp_rank > 0`; engine core init raises instead."""

203 connector = _LegacyConnector()

204

205 with pytest.raises(ValueError, match="does not support PP-disaggregated"):

@@ -216,8 +214,7 @@ def test_engine_passes_handshake_metadata_through_for_pp_aware_connector(

214 monkeypatch: pytest.MonkeyPatch,

215 ) -> None:

216 """A PP-aware connector receives the full `(pp_rank, tp_rank)`-keyed dict

219 - unchanged.

220 - """

217 + unchanged."""

218 metadata_0 = _Metadata()

219 metadata_1 = _Metadata()

220 connector = _PPAwareConnector()

tests/v1/kv_connector/unit/test_hf3fs_client.py

+2 -4

@@ -131,8 +131,7 @@ class TestHf3fsClientResourceManagement:

131

132 def test_close_idempotent_and_handles_cleared(self, tmp_path):

133 """Multiple close() calls must not raise; deregister_fd called exactly

134 - once, all handles set to None, shm.close() invoked.

135 - """

134 + once, all handles set to None, shm.close() invoked."""

135 client, shm_r, shm_w = self._make_client(tmp_path)

136

137 with (

@@ -238,8 +237,7 @@ class TestHf3fsClientResourceManagement:

237

238 def test_constructor_failure_does_not_close_unallocated_shm(self, tmp_path):

239 """Failure before SharedMemory is created must not raise AttributeError

241 - or TypeError from cleanup.

242 - """

240 + or TypeError from cleanup."""

241 with (

242 patch(f"{self._MOD}.HF3FS_AVAILABLE", True),

243 patch(f"{self._MOD}.register_fd"),

tests/v1/kv_connector/unit/test_lmcache_connector.py

+1 -2

@@ -212,8 +212,7 @@ class TestUpdateConnectorOutput:

212 self, mock_connector

213 ):

214 """Test that method returns early when kv_cache_events is not

215 - LMCacheKVEvents.

216 - """

215 + LMCacheKVEvents."""

216 # Create a mock object that is not LMCacheKVEvents

217 fake_events = MagicMock()

218 connector_output = KVConnectorOutput(kv_cache_events=fake_events)

tests/v1/kv_connector/unit/test_mooncake_stats.py

+1 -2

@@ -199,8 +199,7 @@ def _bare_worker() -> MooncakeConnectorWorker:

199 """Construct a MooncakeConnectorWorker skipping __init__ (full init requires

200 a live TransferEngine). Only the attributes touched by the methods under

201 test are populated; role flags and async_zmq_ctx keep __del__'s shutdown

202 - path a no-op.

203 - """

202 + path a no-op."""

203 worker = MooncakeConnectorWorker.__new__(MooncakeConnectorWorker)

204 worker.xfer_stats = MooncakeKVConnectorStats()

205 worker.engine = MagicMock()

tests/v1/kv_connector/unit/test_mooncake_store_connector.py

+3 -6

@@ -77,8 +77,7 @@ def test_store_tp_does_not_override_backend_kv_cache_layout(extra_config):

77

78 def _make_kv_cache_config() -> KVCacheConfig:

79 """Single-group full-attention KVCacheConfig — enough for the connector

80 - constructor's validate() pass.

81 - """

80 + constructor's validate() pass."""

81 spec = FullAttentionSpec(block_size=16, num_kv_heads=8, head_size=64, dtype=None)

82 return KVCacheConfig(

83 num_blocks=4,

@@ -693,8 +692,7 @@ def _poll_lookup(client, req_id, num_tokens=128, block_hashes=(), timeout=5.0):

692

693 def _gated_recv(gate: threading.Event, value: int):

694 """Mock recv side-effect that blocks until ``gate`` is set, so the

696 - executor's lookup can be held pending deterministically.

697 - """

695 + executor's lookup can be held pending deterministically."""

696

697 def recv():

698 gate.wait()

@@ -705,8 +703,7 @@ def _gated_recv(gate: threading.Event, value: int):

703

704 def test_lookup_key_client_non_block_lookup_async():

705 """Non-blocking lookup defers to the executor: None first, hit once the

708 - Future resolves.

709 - """

706 + Future resolves."""

707 vllm_config = _make_vllm_config()

708

709 with patch(

tests/v1/kv_connector/unit/test_mooncake_store_coordinator.py

+18 -36

@@ -33,8 +33,7 @@ def _mamba_align(block_size=32):

33 def _make_coord(groups, hash_block_size, use_eagle=False, retention_interval=None):

34 """Construct a coordinator using the natural LCM of group block sizes as

35 the scheduler block size — mirrors ``resolve_kv_cache_block_sizes`` for

36 - the test fixtures.

37 - """

36 + the test fixtures."""

37 block_sizes = [g.kv_cache_spec.block_size for g in groups]

38 scheduler_block_size = lcm(*block_sizes)

39 return MooncakeStoreCoordinator(

@@ -144,8 +143,7 @@ def test_coordinator_single_full_attention_no_hits():

143

144 def test_coordinator_single_swa_tautological_pool_masks_pre_window():

145 """SWA tautological-pool: hit_length spans full prefix, mask is

147 - tail-window only.

148 - """

146 + tail-window only."""

147 groups = [KVCacheGroupSpec(["L0"], _swa(block_size=16, sliding_window=32))]

148 coord = _make_coord(groups, hash_block_size=16)

149 hs = _hashes(4) # 4 chunks * 16 tokens

@@ -192,8 +190,7 @@ def test_coordinator_hybrid_hole_in_full_clips_both():

190

191 def test_coordinator_group_block_size_double_hash():

192 """Group block_size=32 over hash_block_size=16 hashes: adjacent

195 - hashes merge before pool lookup.

196 - """

193 + hashes merge before pool lookup."""

194 groups = [

195 KVCacheGroupSpec(["L0"], _full(16)),

196 KVCacheGroupSpec(["L1"], _full(32)),

@@ -217,8 +214,7 @@ def test_coordinator_group_block_size_double_hash():

214 def test_coordinator_fine_grained_partial_tail_hit():

215 """K3 shape: FA + mamba-align, block_size=32 over hash_block_size=16. When

216 both groups have the sub-block boundary hash, the reconciled hit lands on

220 - the hash boundary (48), not the block boundary (32).

221 - """

217 + the hash boundary (48), not the block boundary (32)."""

218 groups = [

219 KVCacheGroupSpec(["L0"], _full(32)),

220 KVCacheGroupSpec(["L1"], _mamba_align(32)),

@@ -238,8 +234,7 @@ def test_coordinator_fine_grained_partial_tail_hit():

234

235 def test_coordinator_fine_grained_clips_when_one_group_missing_tail():

236 """If only one group has the sub-block boundary, min-convergence clips the

241 - reconciled hit back to the block boundary (32).

242 - """

237 + reconciled hit back to the block boundary (32)."""

238 groups = [

239 KVCacheGroupSpec(["L0"], _full(32)),

240 KVCacheGroupSpec(["L1"], _mamba_align(32)),

@@ -280,8 +275,7 @@ def test_store_mask_swa_only_window_around_each_lcm_boundary():

275 """Hybrid full-attn(block=32) + SWA(block=8, sw=8). lcm=32. With

276 aligned=64 the SWA group should mark exactly the blocks ending at 32

277 and 64 (i.e. blocks 3 and 7 at block_size=8); the rest can never

283 - participate in any future hit.

284 - """

278 + participate in any future hit."""

279 full = _full(32)

280 swa = _swa(block_size=8, sliding_window=8)

281 groups = [KVCacheGroupSpec(["L0"], full), KVCacheGroupSpec(["L1"], swa)]

@@ -327,8 +321,7 @@ def test_store_mask_swa_does_not_double_scale_under_dcp():

321

322 def test_store_mask_swa_wider_window_covers_more_blocks_per_lcm():

323 """Same hybrid layout but sliding_window=16 (= 2 SWA blocks). Each lcm

330 - boundary should now span two SWA tail blocks.

331 - """

324 + boundary should now span two SWA tail blocks."""

325 full = _full(32)

326 swa = _swa(block_size=8, sliding_window=16)

327 groups = [KVCacheGroupSpec(["L0"], full), KVCacheGroupSpec(["L1"], swa)]

@@ -358,8 +351,7 @@ def test_store_mask_dsv4_5_groups_full_mla_plus_4_swa():

351 """DSV4-shaped: full-MLA(B=256) + 4 SWA groups with B in {64, 64, 4, 8}

352 and varied sliding windows. lcm=256, hash_block_size=4. Two lcm segments

353 (aligned_len=512). Validates that the tile-once strategy produces the

361 - expected per-segment tail-window pattern, repeated.

362 - """

354 + expected per-segment tail-window pattern, repeated."""

355 full_mla = _full(block_size=256)

356 swa_64_sw128 = _swa(block_size=64, sliding_window=128)

357 swa_64_sw512 = _swa(block_size=64, sliding_window=512)

@@ -393,8 +385,7 @@ def test_store_mask_dsv4_5_groups_full_mla_plus_4_swa():

385

386 def test_store_mask_fast_path_all_block_sizes_equal_lcm():

387 """When every non-full-attn group already aligns to lcm_block_size, the

396 - fast path returns all-True without invoking find_longest_cache_hit.

397 - """

388 + fast path returns all-True without invoking find_longest_cache_hit."""

389 full = _full(block_size=64)

390 swa = _swa(block_size=64, sliding_window=128)

391 groups = [KVCacheGroupSpec(["L0"], full), KVCacheGroupSpec(["L1"], swa)]

@@ -407,8 +398,7 @@ def test_store_mask_fast_path_all_block_sizes_equal_lcm():

398

399 def test_store_mask_fast_path_single_attention_group():

400 """Two groups sharing the same SWA spec collapse to one attention group;

410 - no lcm filter applies, every chunk is True.

411 - """

401 + no lcm filter applies, every chunk is True."""

402 swa = _swa(block_size=16, sliding_window=32)

403 groups = [KVCacheGroupSpec(["L0"], swa), KVCacheGroupSpec(["L1"], swa)]

404 coord = _make_coord(groups, hash_block_size=16)

@@ -422,8 +412,7 @@ def test_store_mask_fast_path_single_attention_group():

412

413 def _retention_groups():

414 """Hybrid full-attn(block=32) + SWA(block=8, sw=8); lcm=32. The SWA group

425 - densely keeps one tail block per 32-token boundary.

426 - """

415 + densely keeps one tail block per 32-token boundary."""

416 full = _full(32)

417 swa = _swa(block_size=8, sliding_window=8)

418 return [KVCacheGroupSpec(["L0"], full), KVCacheGroupSpec(["L1"], swa)]

@@ -431,8 +420,7 @@ def _retention_groups():

420

421 def test_store_mask_dense_default_matches_every_lcm_boundary():

422 """retention_interval=None (default) keeps the SWA tail at every lcm

434 - boundary: tokens 32/64/96/128 -> chunks 3/7/11/15.

435 - """

423 + boundary: tokens 32/64/96/128 -> chunks 3/7/11/15."""

424 coord = _make_coord(_retention_groups(), hash_block_size=8)

425 masks = coord.store_mask(128)

426 assert masks[0] is None

@@ -442,8 +430,7 @@ def test_store_mask_dense_default_matches_every_lcm_boundary():

430 def test_store_mask_retention_interval_sparsifies_swa_tails():

431 """retention_interval=64 keeps an SWA tail once per 64-token segment

432 (chunks 7 and 15) instead of every 32 tokens, dropping the mid-segment

445 - boundaries at 32 and 96.

446 - """

433 + boundaries at 32 and 96."""

434 coord = _make_coord(_retention_groups(), hash_block_size=8, retention_interval=64)

435 masks = coord.store_mask(128)

436 assert masks[0] is None # full attn unaffected

@@ -452,8 +439,7 @@ def test_store_mask_retention_interval_sparsifies_swa_tails():

439

440 def test_store_mask_retention_interval_zero_keeps_only_replay_boundary():

441 """retention_interval=0 drops all segment tails; only the latest replay

455 - boundary (capped at num_prompt-1, aligned down to lcm) is retained.

456 - """

442 + boundary (capped at num_prompt-1, aligned down to lcm) is retained."""

443 coord = _make_coord(_retention_groups(), hash_block_size=8, retention_interval=0)

444 # No replay info -> nothing reachable for the SWA group.

445 assert coord.store_mask(128)[1] == [False] * 16

@@ -464,8 +450,7 @@ def test_store_mask_retention_interval_zero_keeps_only_replay_boundary():

450

451 def test_store_mask_retention_interval_keeps_segment_and_replay_tails():

452 """Sparse segment tails (interval=64 -> chunks 7,15) plus the replay

467 - boundary tail (num_prompt=100 -> chunk 11) coexist.

468 - """

453 + boundary tail (num_prompt=100 -> chunk 11) coexist."""

454 coord = _make_coord(_retention_groups(), hash_block_size=8, retention_interval=64)

455 masks = coord.store_mask(128, num_prompt_tokens=100)

456 assert masks[1] == [i in (7, 11, 15) for i in range(16)]

@@ -503,8 +488,7 @@ def test_store_mask_excludes_mamba_groups_lookup_unaffected():

488 nulled/freed; speculative blocks relocate), so the positional normal save

489 must never cover mamba chunks — regardless of retention. Lookups still

490 probe mamba boundaries: their keys come from the pinned snapshot/CoW

506 - hand-off path instead.

507 - """

491 + hand-off path instead."""

492 groups = [

493 KVCacheGroupSpec(["L0"], _full(16)),

494 KVCacheGroupSpec(["L1"], _mamba_align(16)),

@@ -525,8 +509,7 @@ def test_store_mask_excludes_mamba_groups_lookup_unaffected():

509

510 def test_lookup_with_eagle_pops_last_full_attention_block():

511 """Sanity: with use_eagle, find_longest_cache_hit drops the last block.

528 - Pairs with the load_mask test below to lock the round-trip contract.

529 - """

512 + Pairs with the load_mask test below to lock the round-trip contract."""

513 groups = [KVCacheGroupSpec(["L0"], _full(16))]

514 coord = _make_coord(groups, hash_block_size=16, use_eagle=True)

515 hs = _hashes(4)

@@ -567,8 +550,7 @@ def test_load_mask_with_eagle_does_not_double_prune_full_attention():

550 def test_load_mask_with_eagle_hybrid_full_plus_swa():

551 """Hybrid (FullAttn + SWA) with eagle: load_mask must cover every chunk

552 in [0, token_len) for the FullAttn group; SWA group keeps its

570 - tail-window mask.

571 - """

553 + tail-window mask."""

554 groups = [

555 KVCacheGroupSpec(["L0"], _full(16)),

556 KVCacheGroupSpec(["L1"], _swa(16, 32)),

tests/v1/kv_connector/unit/test_mooncake_store_hma_e2e.py

+7 -14

@@ -1,8 +1,7 @@

1 # SPDX-License-Identifier: Apache-2.0

2 # SPDX-FileCopyrightText: Copyright contributors to the vLLM project

3 """End-to-end save->lookup test for MooncakeStoreConnector on a hybrid

4 -(SWA + Full) attention config, using a dict-backed mock store.

5 -"""

4 +(SWA + Full) attention config, using a dict-backed mock store."""

5

6 import sys

7 import threading

@@ -283,8 +282,7 @@ def test_e2e_swa_plus_full_save_then_lookup_hits():

282

283 def test_recv_skips_swa_blocks_before_window():

284 """Producer stored every block for both groups; consumer must only fetch

286 - SWA blocks within the sliding window, not the head.

287 - """

285 + SWA blocks within the sliding window, not the head."""

286 full = FullAttentionSpec(block_size=16, num_kv_heads=8, head_size=64, dtype=None)

287 # sliding_window=32, block_size=16 → 2 contiguous blocks within window.

288 swa = SlidingWindowSpec(

@@ -352,8 +350,7 @@ def test_recv_skips_swa_blocks_before_window():

350

351 def test_chunked_token_database_hash_block_size_smaller_than_block_size():

352 """DSv4-style: hash_block_size=4, group block_size=16 — process_tokens

355 - keys each chunk by its ending fine hash, including a partial tail.

356 - """

353 + keys each chunk by its ending fine hash, including a partial tail."""

354 md = KeyMetadata("m", 0, 0, 0, 0, group_id=3)

355 db = ChunkedTokenDatabase(md, block_size=16, hash_block_size=4)

356 db.set_kv_caches_base_addr([0])

@@ -386,8 +383,7 @@ def test_sub_block_partial_tail_offload_reads_cow_block():

383 """Sub-block prompt (the 900/128/1536 shape, scaled to 12/4/16): the

384 partial tail is offloaded for both groups under the boundary sub-hash. The

385 full-attention block is read from the request block table; the mamba block

389 - is the core-provided CoW target, not block_ids.

390 - """

386 + is the core-provided CoW target, not block_ids."""

387 full = FullAttentionSpec(block_size=16, num_kv_heads=8, head_size=64, dtype=None)

388 mamba = MambaSpec(

389 block_size=16,

@@ -466,8 +462,7 @@ def test_sub_block_partial_tail_offload_reads_cow_block():

462 def test_offload_syncs_event_before_put():

463 """An offload-carrying meta synchronizes its CoW-fence event before the

464 store put reads the blocks, then completes in one pass and drains the

469 - completion counter.

470 - """

465 + completion counter."""

466 full = FullAttentionSpec(block_size=16, num_kv_heads=8, head_size=64, dtype=None)

467 mamba = MambaSpec(

468 block_size=16,

@@ -538,8 +533,7 @@ def test_sub_block_partial_tail_offload_covers_smaller_group_blocks():

533 offload must persist every FA block up to the boundary — the normal save

534 floors to the lcm, so those blocks are otherwise never written and the

535 consumer's per-group lookup would miss. The mamba boundary block still

541 - reads the core-provided CoW target.

542 - """

536 + reads the core-provided CoW target."""

537 full = FullAttentionSpec(block_size=4, num_kv_heads=8, head_size=64, dtype=None)

538 mamba = MambaSpec(

539 block_size=16,

@@ -810,8 +804,7 @@ def test_ring_scratch_group_is_never_stored_and_does_not_block_hits():

804 """A per-request ring group (CircularBufferSpec, capacity as block_size)

805 sits beside the paged group: worker setup tolerates its block size (the

806 DeepSeek-V4.1 compressor ring is 8 rows), saving a request stores paged

813 - blocks only, and lookup hits on the paged group alone.

814 - """

807 + blocks only, and lookup hits on the paged group alone."""

808 full = FullAttentionSpec(block_size=16, num_kv_heads=8, head_size=64, dtype=None)

809 ring = CircularBufferSpec(

810 block_size=8, num_kv_heads=1, head_size=64, head_size_v=0, dtype=torch.uint8

tests/v1/kv_connector/unit/test_mooncake_store_scheduler.py

+2 -4

@@ -536,8 +536,7 @@ def _make_pending_load_unfinished_request(

536

537 def _make_pending_load_scheduler_output() -> SimpleNamespace:

538 """scheduler_output for a step where req-0 is parked on a pending load

539 - (not in scheduled_new_reqs or scheduled_cached_reqs).

540 - """

539 + (not in scheduled_new_reqs or scheduled_cached_reqs)."""

540 return SimpleNamespace(

541 finished_req_ids=set(),

542 preempted_req_ids=set(),

@@ -713,8 +712,7 @@ def test_running_request_not_in_resumed_req_ids_appends_blocks():

712 def test_resumed_request_in_resumed_req_ids_replaces_blocks():

713 """A request the scheduler marks resumed gets the FULL block table in

714 new_block_ids and must REPLACE the tracker's blocks (not append), even if

716 - a stale tracker from before preemption is still present.

717 - """

715 + a stale tracker from before preemption is still present."""

716 scheduler = _make_bare_scheduler()

717 _make_resumed_unfinished_request(

718 scheduler,

tests/v1/kv_connector/unit/test_mooncake_store_worker.py

+20 -40

@@ -432,8 +432,7 @@ def _patch_worker_runtime(

432

433 def test_pool_key_to_string_without_prefix_is_unchanged():

434 """Default (empty) cache_prefix keeps keys byte-identical to the

435 - historical unprefixed format so existing deployments keep their hits.

436 - """

435 + historical unprefixed format so existing deployments keep their hits."""

436 key = PoolKey(KeyMetadata("test-model", 0, 0, 0, 0), "deadbeef")

437 assert (

438 key.to_string() == "test-model@tp_rank:0@pcp0@dcp0@pp_rank:0@group:0@deadbeef"

@@ -574,8 +573,7 @@ def test_tp_shared_receiving_reads_each_local_store_shard():

573

574 def test_pool_key_cache_prefix_namespaces_and_disambiguates():

575 """A non-empty cache_prefix is prepended, and two instances with

577 - different prefixes never collide on identical block hashes.

578 - """

576 + different prefixes never collide on identical block hashes."""

577 md_a = KeyMetadata("test-model", 0, 0, 0, 0, cache_prefix="depA")

578 md_b = KeyMetadata("test-model", 0, 0, 0, 0, cache_prefix="depB")

579

@@ -591,8 +589,7 @@ def test_pool_key_cache_prefix_namespaces_and_disambiguates():

589

590 def test_default_local_buffer_size_matches_pr40900():

591 """PR-40900 shipped a 4 GiB default for local_buffer_size; the dual-mode

594 - patch preserves it (and the JSON key) so unchanged PR-40900 configs work.

595 - """

592 + patch preserves it (and the JSON key) so unchanged PR-40900 configs work."""

593 assert worker.DEFAULT_LOCAL_BUFFER_SIZE == 4 * 1024**3

594

595

@@ -1014,8 +1011,7 @@ def test_store_sending_thread_skips_null_sparse_group_blocks():

1011

1012 def test_partial_tail_offload_skips_cap_omitted_mamba_group():

1013 """A Mamba group omitted by the handoff cap must not fall back to the

1017 - connector's positional block table.

1018 - """

1014 + connector's positional block table."""

1015 from vllm.v1.kv_cache_interface import (

1016 FullAttentionSpec,

1017 KVCacheGroupSpec,

@@ -1143,8 +1139,7 @@ def test_partial_tail_put_failure_activates_pressure_gate():

1139 def test_normal_save_excludes_mamba_group_and_null_blocks():

1140 """The positional normal save must never cover mamba chunks (align-mode

1141 block tables are not append-only) and must skip chunks whose block is the

1146 - reserved null block, for any group.

1147 - """

1142 + reserved null block, for any group."""

1143 from vllm.v1.kv_cache_interface import (

1144 FullAttentionSpec,

1145 KVCacheGroupSpec,

@@ -1205,8 +1200,7 @@ def test_normal_save_excludes_mamba_group_and_null_blocks():

1200 def test_block_aligned_snapshot_offload_uses_provided_block():

1201 """A block-aligned hand-off (sparse-retention mamba checkpoint) uploads

1202 exactly the core-pinned block under the boundary-end hash key; the

1208 - positional block table entry for that chunk must be ignored.

1209 - """

1203 + positional block table entry for that chunk must be ignored."""

1204 store = MagicMock()

1205 store.batch_is_exist.side_effect = lambda keys: [0] * len(keys)

1206 store.batch_put_from_multi_buffers.side_effect = lambda keys, *a: [256] * len(keys)

@@ -1234,8 +1228,7 @@ def test_block_aligned_snapshot_offload_uses_provided_block():

1228 def test_mixed_snapshot_and_sub_block_offloads():

1229 """A retention snapshot and the prompt-end sub-block CoW tail can arrive

1230 in one hand-off; the sub-block path covers FA gap blocks but reads the

1237 - mamba boundary only from the CoW block, never positionally.

1238 - """

1231 + mamba boundary only from the CoW block, never positionally."""

1232 store = MagicMock()

1233 store.batch_is_exist.side_effect = lambda keys: [0] * len(keys)

1234 store.batch_put_from_multi_buffers.side_effect = lambda keys, *a: [256] * len(keys)

@@ -1275,8 +1268,7 @@ def test_mixed_snapshot_and_sub_block_offloads():

1268

1269 def test_snapshot_offload_skips_null_handoff_block():

1270 """A hand-off the core could not materialize (null block) carries no

1278 - committed state; persisting it would poison the boundary key.

1279 - """

1271 + committed state; persisting it would poison the boundary key."""

1272 store = MagicMock()

1273 thread = _make_partial_tail_send_thread(store)

1274

@@ -1604,8 +1596,7 @@ def test_store_sending_thread_passes_replicate_config_when_preferred_segment_set

1596 def test_store_sending_thread_passes_default_replicate_config_when_no_preferred_segment(): # noqa: E501

1597 """Without a preferred_segment the SendingThread still forwards a

1598 (default-constructed) ReplicateConfig so the C++ side always sees a

1607 - well-defined config object.

1608 - """

1599 + well-defined config object."""

1600 store = MagicMock()

1601 store.batch_is_exist.side_effect = lambda keys: [0] * len(keys)

1602 store.batch_put_from_multi_buffers.return_value = [256, 256]

@@ -2115,8 +2106,7 @@ def test_recv_thread_stops_after_first_failing_disk_offload_sub_batch():

2106

2107 def test_recv_thread_skips_split_when_budget_holds_all_keys():

2108 """PR-36 removed the count-based split trigger; with budget for 3 keys,

2118 - all three should be requested in a single call.

2119 - """

2109 + all three should be requested in a single call."""

2110 store = MagicMock()

2111 store.batch_get_into_multi_buffers.return_value = [256, 256, 256]

2112 thread = _make_store_recving_thread(

@@ -2330,8 +2320,7 @@ def test_requester_worker_init_skips_disk_budget_when_offload_disabled(

2320 monkeypatch,

2321 ):

2322 """enable_offload=False zeroes out the disk budget so we don't generate

2333 - redundant owner GET-RPCs.

2334 - """

2323 + redundant owner GET-RPCs."""

2324 store = MagicMock()

2325 store.setup.return_value = 0

2326 _install_fake_mooncake(monkeypatch, store)

@@ -3006,8 +2995,7 @@ def test_requester_worker_group_semantics_string_true_enables(

2995

2996 def test_store_sending_thread_clamps_token_len_to_lcm():

2997 """Partial chunks past the last lcm boundary aren't stored — cache hits

3009 - are always lcm-aligned (mirrors HybridKVCacheCoordinator).

3010 - """

2998 + are always lcm-aligned (mirrors HybridKVCacheCoordinator)."""

2999 store = MagicMock()

3000 store.batch_is_exist.return_value = [0, 0]

3001 store.batch_put_from_multi_buffers.return_value = [256, 256]

@@ -3032,8 +3020,7 @@ def test_store_sending_thread_clamps_token_len_to_lcm():

3020

3021 def test_store_sending_thread_skips_when_token_len_below_lcm():

3022 """Requests shorter than lcm_block_size cannot produce any aligned chunk,

3035 - so neither the existence check nor the put should be issued.

3036 - """

3023 + so neither the existence check nor the put should be issued."""

3024 from vllm.v1.kv_cache_interface import FullAttentionSpec, KVCacheGroupSpec

3025

3026 store = MagicMock()

@@ -3823,8 +3810,7 @@ def test_lookup_full_hit_with_eagle_pops_once_not_twice():

3810 """Eagle already leaves the last block for the drafter, so a

3811 full-prompt re-derivation must never fire for eagle-governed hits:

3812 firing would anchor the search one block lower and pop a second

3826 - block, regressing the hit by an extra producer boundary.

3827 - """

3813 + block, regressing the hit by an extra producer boundary."""

3814 worker = _make_bare_worker(block_size=16)

3815 worker.coord = mooncake_store_worker.MooncakeStoreCoordinator(

3816 worker._kv_cache_groups,

@@ -3981,8 +3967,7 @@ def test_lookup_full_hit_swa_degrades_when_no_stored_boundary_is_usable():

3967 the full hit into 48, whose SWA window needs the never-written block 1,

3968 so every load failed and the recompute re-entered the same lookup. The

3969 re-derivation must report that no stored boundary below the request end

3984 - is usable.

3985 - """

3970 + is usable."""

3971 from vllm.v1.kv_cache_interface import KVCacheGroupSpec, SlidingWindowSpec

3972

3973 worker = _make_bare_worker(block_size=16)

@@ -4004,8 +3989,7 @@ def test_lookup_full_hit_swa_degrades_when_no_stored_boundary_is_usable():

3989 def test_lookup_swa_single_group_returns_full_when_tail_window_present():

3990 """Single-SWA, sliding_window=32 (= 2 blocks): producer stored only the

3991 tail. Coordinator-driven lookup returns full prefix even though the

4007 - pre-window blocks are absent.

4008 - """

3992 + pre-window blocks are absent."""

3993 from vllm.v1.kv_cache_interface import KVCacheGroupSpec, SlidingWindowSpec

3994

3995 worker = _make_bare_worker(block_size=16)

@@ -4401,8 +4385,7 @@ def test_config_defaults_to_embedded():

4385

4386 def test_config_pr40900_unchanged(tmp_path):

4387 """A literal PR-40900 config (no mode, no enable_offload, no preferred_segment)

4404 - parses without raising and resolves to embedded mode.

4405 - """

4388 + parses without raising and resolves to embedded mode."""

4389 config_path = _write_mooncake_config(

4390 tmp_path,

4391 {

@@ -4510,8 +4493,7 @@ def test_config_zero_local_buffer():

4493 def test_topology_standalone_store_with_disk_offload(tmp_path, monkeypatch):

4494 """standalone-store + disk: global_segment_size=0, enable_offload=True,

4495 preferred_segment set. Assert setup() positional args, ReplicateConfig

4513 - wiring, and that the disk-offload buffer budget is allocated.

4514 - """

4496 + wiring, and that the disk-offload buffer budget is allocated."""

4497 store = MagicMock()

4498 store.setup.return_value = 0

4499 fake_replicate_config_cls = _install_fake_mooncake(monkeypatch, store)

@@ -4560,8 +4542,7 @@ def test_topology_standalone_store_with_disk_offload(tmp_path, monkeypatch):

4542 def test_topology_embedded_cpu_only(tmp_path, monkeypatch):

4543 """Embedded + CPU-only: no mode key (defaults to embedded),

4544 global_segment_size>0, enable_offload absent, no preferred_segment.

4563 - This is the PR-40900 baseline recipe.

4564 - """

4545 + This is the PR-40900 baseline recipe."""

4546 store = MagicMock()

4547 store.setup.return_value = 0

4548 fake_replicate_config_cls = _install_fake_mooncake(monkeypatch, store)

@@ -4750,8 +4731,7 @@ def test_store_worker_close_swallows_store_errors():

4731 def test_blob_block_hashes_wire_roundtrip():

4732 """The lookup wire format sends a ``hash_len`` frame plus the raw hashes

4733 concatenated back-to-back; the server rebuilds them through a zero-copy

4753 - ``BlobBlockHashes`` view over the frame buffer.

4754 - """

4734 + ``BlobBlockHashes`` view over the frame buffer."""

4735 hashes = [BlockHash(bytes([i]) * 16) for i in range(5)]

4736 hash_len = len(hashes[0])

4737

tests/v1/kv_connector/unit/test_moriio_connector.py

+7 -14

@@ -692,8 +692,7 @@ def test_moriio_handshake_returns_metadata(mock_parallel_groups):

692 def test_resolve_host_ip_prefers_extra_config():

693 """An explicit ``host_ip`` in kv_connector_extra_config overrides get_ip()

694 (so an external router can advertise a routable/internal address); an

695 - absent or empty value falls back to get_ip().

696 - """

695 + absent or empty value falls back to get_ip()."""

696 assert resolve_host_ip({"host_ip": "10.0.0.7"}) == "10.0.0.7"

697

698 fallback = get_ip()

@@ -747,8 +746,7 @@ def _read_scheduler(

746

747 def test_hma_blocks_per_sw_two_groups():

748 """A Full + SlidingWindow config runs with HMA and computes block budgets

750 - correctly

751 - """

749 + correctly"""

750 scheduler = _read_scheduler(_make_hybrid_kv_cache_config())

751 assert scheduler._is_hma_required is True

752 # cdiv(32, 16) + 1 == 3 for the sliding-window group, 0 for full attention.

@@ -766,8 +764,7 @@ def test_hma_blocks_per_sw_two_groups():

764 def test_is_hma_required(swa_enabled, disable_hma, expected_is_hma):

765 """_is_hma_required tracks both the KV cache groups and the

766 --disable-hybrid-kv-cache-manager flag. When HMA is off,

769 - get_exchange_clipped_blocks must be a no-op.

770 - """

767 + get_exchange_clipped_blocks must be a no-op."""

768 config = (

769 _make_hybrid_kv_cache_config() if swa_enabled else _make_test_kv_cache_config()

770 )

@@ -780,8 +777,7 @@ def test_is_hma_required(swa_enabled, disable_hma, expected_is_hma):

777

778 def test_non_sliding_window_hybrid_is_rejected():

779 """A hybrid group that is not sliding-window (e.g. chunked-local

783 - attention) must fail closed rather than be silently mistransferred.

784 - """

780 + attention) must fail closed rather than be silently mistransferred."""

781 full_spec = FullAttentionSpec(

782 block_size=16, num_kv_heads=4, head_size=64, dtype=torch.float16

783 )

@@ -815,8 +811,7 @@ def test_non_sliding_window_hybrid_is_rejected():

811

812 def test_token_count_basis_uses_full_attention_group():

813 """Chunked-prefill token counting must use an unclipped full-attention

818 - group (blocks_per_sw == 0), not a clipped sliding-window group.

819 - """

814 + group (blocks_per_sw == 0), not a clipped sliding-window group."""

815 scheduler = _read_scheduler(_make_hybrid_kv_cache_config())

816 # Group 0 is the full-attention group

817 assert scheduler._full_attn_group_idx == 0

@@ -825,8 +820,7 @@ def test_token_count_basis_uses_full_attention_group():

820

821 def test_get_exchange_clipped_blocks_clips_only_sw_group():

822 """get_exchange_clipped_blocks keeps the full attn group intact and clips the

828 - sliding-window group to its window tail.

829 - """

823 + sliding-window group to its window tail."""

824 scheduler = _read_scheduler(_make_hybrid_kv_cache_config())

825 full = [10, 11, 12, 13, 14]

826 sw = [20, 21, 22, 23, 24]

@@ -837,8 +831,7 @@ def test_get_exchange_clipped_blocks_clips_only_sw_group():

831

832 def test_metadata_hma_block_ids_preserved_per_group():

833 """add_new_req stores per-group (BlockIds) block lists unchanged for both

840 - read and write, so the hybrid group structure is retained.

841 - """

834 + read and write, so the hybrid group structure is retained."""

835 metadata = MoRIIOConnectorMetadata()

836

837 # Assume:

tests/v1/kv_connector/unit/test_moriio_routing_fairness.py

+2 -4

@@ -110,8 +110,7 @@ def make_meta(*, p_tp: int, p_dp: int, remote_dp_rank: int, host: str = "phost0"

110 def build_decode_workers(cfg: PDConfig) -> list:

111 """Decode workers that issue reads for one prefill instance. A TP decode

112 instance reads from every tp rank; a DP decode instance from each dp-rank

113 - worker. Reused across requests so per-worker round-robin state advances.

114 - """

113 + worker. Reused across requests so per-worker round-robin state advances."""

114 if cfg.d_tp > 1:

115 return [

116 make_decode_worker(world_size=cfg.d_tp, tp_rank=r, dp_rank=0)

@@ -126,8 +125,7 @@ def prefill_target_multiset(cfg: PDConfig, rounds: int) -> Counter:

125 """Drive the REAL _resolve_read_source over a fair input stream; tally which

126 prefill GPU each read targets. The only modelled assumption is the proxy

127 contract -- each owner dp-rank delivered equally (``for owner_dp in

129 - range(p_dp)``); no proxy algorithm is reproduced.

130 - """

128 + range(p_dp)``); no proxy algorithm is reproduced."""

129 workers = build_decode_workers(cfg)

130 hits: Counter = Counter()

131 for _ in range(rounds):

tests/v1/kv_connector/unit/test_multi_connector.py

+1 -2

@@ -555,8 +555,7 @@ class TestMultiConnectorStats:

555

556 def test_build_kv_connector_stats_reconstructs_nixl_stats(self):

557 """Test that NixlConnector stats are properly reconstructed with

558 - correct data.

559 - """

558 + correct data."""

559 serialized_data = {

560 "NixlConnector": {

561 "transfer_duration": [1.5, 2.3],

tests/v1/kv_connector/unit/test_nixl_connector.py

+2 -4

@@ -318,8 +318,7 @@ def test_abort_immediately_remote_prefill_enqueues_empty_recv():

318 """A remote-prefill request added with abort_immediately=True should

319 be added to the scheduler's waiting queue then immediately aborted, so the

320 NIXL connector's request_finished hook enqueues an empty recv to notify

321 - the prefill instance to free its blocks.

322 - """

321 + the prefill instance to free its blocks."""

322 from vllm.v1.request import RequestStatus

323

324 scheduler = create_scheduler(create_vllm_config())

@@ -3146,8 +3145,7 @@ def test_split_read_failure_defers_report_until_last_handle(

3145 request — nor invalidate its blocks — while the sibling xfer is still in

3146 flight: a posted READ cannot be aborted and would DMA into blocks the

3147 scheduler could free and reuse. The report happens exactly once, when the

3149 - last handle is terminal.

3150 - """

3148 + last handle is terminal."""

3149 request_id = "split_read_partial_failure"

3150 err_handle, live_handle = 11, 22

3151 connector, worker, wrapper = _make_split_read_connector(

tests/v1/kv_connector/unit/test_nixl_connector_hma.py

+17 -34

@@ -281,8 +281,7 @@ def test_update_state_after_alloc_tracks_cached_blocks_per_group():

281 """Hybrid SWA+FA requests can have different prefix-cache-hit counts per

282 KV cache group. Each group's count must be tracked independently rather

283 than collapsed to a single scalar, or DCP read-phase alignment

284 - (local_num_computed_blocks) would misalign one of the groups.

285 - """

284 + (local_num_computed_blocks) would misalign one of the groups."""

285 from vllm.distributed.kv_transfer.kv_connector.v1.nixl.pull_scheduler import (

286 NixlPullConnectorScheduler,

287 )

@@ -336,8 +335,7 @@ def test_full_local_hit_is_not_awaited_by_the_scheduler():

335 owed to the producer -- so the worker must not report it in

336 finished_recving. Doing so trips `assert RequestStatus.is_finished` in

337 _update_from_kv_xfer_finished, which only tolerates a finished recv for a

339 - request that is parked or already done.

340 - """

338 + request that is parked or already done."""

339 from vllm.distributed.kv_transfer.kv_connector.v1.nixl.pull_scheduler import (

340 NixlPullConnectorScheduler,

341 )

@@ -442,8 +440,7 @@ def test_apply_dcp_prefix_caching_matches_global_logical_positions(

440 """_apply_dcp_prefix_caching must operate on logical block IDs (DCP

441 interleaves at logical-block granularity) and always return equal-length

442 local/remote slices, so a later, unconditional _apply_prefix_caching

445 - trim is a guaranteed no-op rather than double-processing the read.

446 - """

443 + trim is a guaranteed no-op rather than double-processing the read."""

444 from vllm.distributed.kv_transfer.kv_connector.v1.nixl.worker import (

445 NixlConnectorWorker,

446 )

@@ -842,8 +839,7 @@ def test_apply_prefix_caching_ssm_prefix_cache_hit(

839 def test_apply_prefix_caching_ssm_unpairable_slots_rejected():

840 """Local SSM slots can only exceed the remote ones by the position D

841 recomputes itself. A larger excess means the lists aren't

845 - position-aligned: fail loudly rather than transfer into wrong slots.

846 - """

842 + position-aligned: fail loudly rather than transfer into wrong slots."""

843 from vllm.distributed.kv_transfer.kv_connector.v1.nixl.worker import (

844 NixlConnectorWorker,

845 )

@@ -1159,8 +1155,7 @@ def test_nixl_region_sort_preserves_pipeline_layer_order():

1155 @pytest.mark.cpu_test

1156 def test_get_block_descs_ids_hybrid_ssm():

1157 """Test _compute_desc_ids uses per-group strides for hybrid

1162 - FA+SSM when ratio=1 (no kernel block size mismatch).

1163 - """

1158 + FA+SSM when ratio=1 (no kernel block size mismatch)."""

1159 from vllm.v1.kv_cache_interface import FullAttentionSpec, MambaSpec

1160

1161 worker = _make_mock_worker_for_desc_ids(

@@ -1244,8 +1239,7 @@ def test_get_block_descs_ids_uses_per_region_pool_capacity():

1239 @pytest.mark.cpu_test

1240 def test_get_block_descs_ids_kernel_block_mismatch():

1241 """Test _compute_desc_ids uses different strides for FA

1247 - (kernel blocks) vs SSM (logical blocks) when ratio > 1.

1248 - """

1242 + (kernel blocks) vs SSM (logical blocks) when ratio > 1."""

1243 from vllm.v1.kv_cache_interface import FullAttentionSpec, MambaSpec

1244

1245 ratio = 4

@@ -1276,8 +1270,7 @@ def test_get_block_descs_ids_kernel_block_mismatch():

1270 def test_get_block_descs_ids_hetero_block_size_hybrid():

1271 """With a block-size ratio, FA desc ids are ratio-expanded while SSM

1272 desc ids keep the unexpanded logical stride (state blocks are never

1279 - sub-split).

1280 - """

1273 + sub-split)."""

1274 from vllm.v1.kv_cache_interface import FullAttentionSpec, MambaSpec

1275

1276 worker = _make_mock_worker_for_desc_ids(

@@ -1316,8 +1309,7 @@ def _bind_worker_method(worker, name):

1309 @pytest.mark.cpu_test

1310 def test_map_block_ids_for_block_size_ratio_hybrid():

1311 """Attention groups expand to remote granularity and clip to the remote

1319 - coverage; mamba state blocks pass through 1:1.

1320 - """

1312 + coverage; mamba state blocks pass through 1:1."""

1313 from unittest.mock import MagicMock

1314

1315 from vllm.distributed.kv_transfer.kv_connector.v1.nixl.worker import (

@@ -1348,8 +1340,7 @@ def test_map_block_ids_for_block_size_ratio_hybrid():

1340 @pytest.mark.cpu_test

1341 def test_post_process_zeroes_untransferred_tail():

1342 """The untransferred sub-blocks of the last local block are zeroed on

1351 - receive; mamba state caches are untouched by the attention permute.

1352 - """

1343 + receive; mamba state caches are untouched by the attention permute."""

1344 from unittest.mock import MagicMock

1345

1346 from vllm.distributed.kv_transfer.kv_connector.v1.nixl.worker import (

@@ -1392,8 +1383,7 @@ def test_post_process_zeroes_untransferred_tail():

1383 @pytest.mark.cpu_test

1384 def test_nixl_metadata_hybrid_ssm_block_ids():

1385 """Test NixlConnectorMetadata correctly stores block IDs for FA + SSM

1395 - groups with different block counts (kernel mismatch active).

1396 - """

1386 + groups with different block counts (kernel mismatch active)."""

1387 from vllm.distributed.kv_transfer.kv_connector.v1.nixl.metadata import (

1388 NixlConnectorMetadata,

1389 )

@@ -1474,8 +1464,7 @@ def _make_fake_kv_cache_manager():

1464 @pytest.mark.cpu_test

1465 def test_zeroing_block_ids_cover_only_loaded_attention_blocks():

1466 """Only zero-recorded (attention) groups contribute, sliced to the

1477 - externally-loaded token range; Mamba state blocks are never zeroed.

1478 - """

1467 + externally-loaded token range; Mamba state blocks are never zeroed."""

1468 manager = _make_fake_kv_cache_manager()

1469

1470 # Tokens [0, 16) are locally cached; the load covers tokens [16, 56).

@@ -1503,8 +1492,7 @@ def test_scheduler_filters_connector_loaded_blocks_from_zeroing():

1492 @pytest.mark.cpu_test

1493 def test_failed_load_rezeroes_unwritten_skipped_blocks():

1494 """A failed async load leaves zeroing-skipped blocks unwritten beyond

1506 - the valid prefix; they must be zeroed before local recompute.

1507 - """

1495 + the valid prefix; they must be zeroed before local recompute."""

1496 from unittest.mock import MagicMock

1497

1498 from vllm.v1.core.sched.scheduler import Scheduler

@@ -1964,8 +1952,7 @@ def test_logical_to_kernel_block_ids_with_remote_ratio(

1952 def test_exchange_clipped_blocks_ssm_single_state():

1953 """In single-state cache modes, SSM lists are reduced to the running

1954 state slot: speculative scratch slots, null placeholders and the previous

1967 - step's state carry nothing. Attention groups pass through untouched.

1968 - """

1955 + step's state carry nothing. Attention groups pass through untouched."""

1956 sched = make_nixl_scheduler(has_mamba=True, is_hma_required=True)

1957 sched.blocks_per_sw = [0, 0]

1958 sched._ssm_spec_blocks = [None, 2]

@@ -1992,8 +1979,7 @@ def test_exchange_clipped_blocks_ssm_single_state():

1979 @pytest.mark.cpu_test

1980 def test_exchange_clipped_blocks_ssm_positional_states():

1981 """In "all" mode every position holds a state, so only the speculative

1995 - slots go; placeholders stay to keep the list position-indexed.

1996 - """

1982 + slots go; placeholders stay to keep the list position-indexed."""

1983 sched = make_nixl_scheduler(has_mamba=True, is_hma_required=True)

1984 sched.blocks_per_sw = [0, 0]

1985 sched._ssm_spec_blocks = [None, 2]

@@ -2010,8 +1996,7 @@ def _make_hybrid_mla_kv_cache_config(num_blocks: int = 4):

1996 """KimiLinear-shaped config: one MLA group and two KDA (GDN-typed

1997 MambaSpec) groups whose layers share the same HMA tensors, with a

1998 mamba-aligned unified page and an MLA kernel block smaller than the

2013 - logical block.

2014 - """

1999 + logical block."""

2000 from vllm.v1.attention.backends.registry import MambaAttentionBackendEnum

2001 from vllm.v1.kv_cache_interface import (

2002 KVCacheConfig,

@@ -2151,8 +2136,7 @@ def test_register_kv_caches_hybrid_mla_dual_purpose_regions():

2136 """Hybrid MLA+KDA registration: HMA tensors shared by both layer types

2137 must be flagged as MLA regions even when a KDA layer registers them

2138 first, expose TP-independent kernel-granularity block lens, and build

2154 - FA + mamba descriptors for every region.

2155 - """

2139 + FA + mamba descriptors for every region."""

2140 from unittest.mock import MagicMock

2141

2142 from vllm.config import set_current_vllm_config

@@ -2253,8 +2237,7 @@ def test_register_kv_caches_hybrid_mla_dual_purpose_regions():

2237 def test_push_write_hybrid_mla_replicates_attention():

2238 """Hybrid MLA+SSM push with P_TP < D_TP: attention blocks must be

2239 written to every covered D rank (replicated MLA latent) while SSM state

2256 - is written per-rank through the split handles.

2257 - """

2240 + is written per-rank through the split handles."""

2241 import threading

2242 from collections import defaultdict

2243 from unittest.mock import MagicMock

tests/v1/kv_connector/unit/test_nixl_desc_geometry.py

+10 -20

@@ -24,8 +24,7 @@ from .utils import create_vllm_config

24

25 class _RecordingNixl:

26 """Minimal NIXL wrapper stand-in that records descriptor lists and

27 - prepared transfers so tests can resolve desc ids to byte ranges.

28 - """

27 + prepared transfers so tests can resolve desc ids to byte ranges."""

28

29 def __init__(self, *args, **kwargs):

30 self.dlists: dict[int, np.ndarray] = {}

@@ -594,8 +593,7 @@ def _register_remote_agents(worker, metadata, tp_size):

593

594 def _owned_byte_ranges(worker, group_logical_ids):

595 """Byte ranges owned by a request: for each HMA region tensor, every

597 - logical block id of every group maps to one unified page.

598 - """

596 + logical block id of every group maps to one unified page."""

597 unified_page = worker._test_unified_page

598 bases = [t.data_ptr() for t in worker._test_tensors]

599 owned = []

@@ -633,8 +631,7 @@ def test_hetero_ppl_multi_read_writes_stay_within_request_blocks():

631 (kernel 4, ppl=3) vs remote (P, TP2) logical blocks of 8 tokens (ppl=2),

632 equal kernel pages, tp_ratio=-2 multi-read with replicated MLA and

633 TP-sharded KDA state. Every local descriptor of the request's reads must

636 - stay within its own blocks.

637 - """

634 + stay within its own blocks."""

635 from vllm.distributed.kv_transfer.kv_connector.v1.nixl.metadata import (

636 NixlConnectorMetadata,

637 )

@@ -701,8 +698,7 @@ def _resolve(

698 """Resolve a desc id to (region, kind, token_start) where kind is 'attn'

699 (desc-page sized, sub-block-aligned, in the request's attention blocks)

700 or 'mamba'. token_start is the request-relative token offset, so local

704 - and remote are comparable even when their kernel blocks differ in size.

705 - """

701 + and remote are comparable even when their kernel blocks differ in size."""

702 addr, length, _ = (int(x) for x in desc_arr[int(idx)])

703 for region, base in enumerate(bases):

704 off = addr - base

@@ -896,8 +892,7 @@ def _run_hetero_case(

892 def test_hetero_ppl_token_alignment_sweep(local_block, remote_block, num_tokens):

893 """Sweep prompt lengths across block-boundary residues for several

894 hetero-ppl geometries; assert neighbor-safety, token alignment, and

899 - coverage of every transferred kernel block.

900 - """

895 + coverage of every transferred kernel block."""

896 _run_hetero_case(

897 local_block, kernel=4, remote_block=remote_block, num_tokens=num_tokens

898 )

@@ -916,8 +911,7 @@ def test_hetero_ppl_with_block_size_ratio(num_tokens):

911 2). The transfer is clipped at remote sub-block granularity by the

912 pairing and front-trimmed by _apply_prefix_caching, so the

913 untransferred tail can span both a partial block and whole blocks —

919 - the case each of the two former zeroing paths handled only half of.

920 - """

914 + the case each of the two former zeroing paths handled only half of."""

915 _run_hetero_case(

916 local_block=24,

917 kernel=8,

@@ -966,8 +960,7 @@ def test_mla_hybrid_large_ppl_geometry(num_tokens):

960 """KimiLinear-scale MLA-hybrid geometry (TP8 prefill -> TP1 decode):

961 decode (local) logical block 5760 / kernel 64 (ppl=90), prefill

962 (remote) logical block 768 (ppl=12), tp_ratio=-8 multi-read with

969 - replicated MLA and 8-way TP-sharded KDA state.

970 - """

963 + replicated MLA and 8-way TP-sharded KDA state."""

964 _run_hetero_case(

965 local_block=5760,

966 kernel=64,

@@ -981,8 +974,7 @@ def test_mla_hybrid_large_ppl_geometry(num_tokens):

974 def test_mismatched_mla_kernel_page_rejected_for_mla_hybrid():

975 """The MLA per-token page is TP-independent, so kernel block lengths

976 differing by anything other than the block-size ratio must fail the

984 - handshake loudly rather than transfer at mismatched geometry.

985 - """

977 + handshake loudly rather than transfer at mismatched geometry."""

978 worker = _make_mla_hybrid_worker(

979 local_block_size=12, kernel_block_size=4, num_logical_blocks=8

980 )

@@ -1166,8 +1158,7 @@ def test_csa_linear_registration_discovers_shared_regions():

1158 def test_csa_linear_scratch_descs_follow_the_pages_they_overlay():

1159 """The scratch ring is addressed by the regions it actually registered in,

1160 not by inferring that it must sit on the MLA pages: overlaying the main KV

1169 - pages instead moves its descriptors with it.

1170 - """

1161 + pages instead moves its descriptors with it."""

1162 worker = _make_csa_linear_ple_worker(scratch_aliases="main_kv")

1163

1164 # Same MLA regions as before, but the ring now lives in the non-MLA ones.

@@ -1301,8 +1292,7 @@ def test_csa_linear_remote_ple_is_copied_whole():

1292

1293 def _make_ring_worker():

1294 """Paged MLA group plus a per-layer ring group, no Mamba: the shape of a

1304 - model whose sliding-window KV lives in per-request rings.

1305 - """

1295 + model whose sliding-window KV lives in per-request rings."""

1296 from unittest.mock import MagicMock

1297

1298 from vllm.config import set_current_vllm_config

tests/v1/kv_connector/unit/test_nixl_push_connector.py

+35 -70

@@ -115,8 +115,7 @@ class _BlocksMock:

115

116 def _stub_sw_clipping(scheduler) -> None:

117 """Make ``get_exchange_clipped_blocks`` a passthrough so tests don't

118 - need the full sliding-window machinery.

119 - """

118 + need the full sliding-window machinery."""

119 scheduler.get_exchange_clipped_blocks = lambda block_ids, clip_ssm=True: block_ids

120

121

@@ -174,8 +173,7 @@ class TestPushScheduler:

173

174 def test_p_side_request_finished_stages_blocks(self):

175 """P scheduler pushes blocks into both _finished_request_blocks (lease)

177 - and _newly_finished_push_blocks (metadata for next step).

178 - """

176 + and _newly_finished_push_blocks (metadata for next step)."""

177 sched = make_nixl_push_scheduler()

178 _stub_sw_clipping(sched)

179

@@ -194,8 +192,7 @@ class TestPushScheduler:

192

193 def test_build_connector_meta_drains_both_sides(self):

194 """meta.push_registrations and meta.push_finished_blocks are filled

197 - from the staging dicts and the staging dicts are cleared.

198 - """

195 + from the staging dicts and the staging dicts are cleared."""

196 sched = make_nixl_push_scheduler()

197 _stub_sw_clipping(sched)

198

@@ -290,8 +287,7 @@ class TestPushScheduler:

287

288 def test_registration_watchdog_expires(self, caplog):

289 """Stale D registrations whose deadline has passed are dropped at

293 - ``build_connector_meta`` time.

294 - """

290 + ``build_connector_meta`` time."""

291 # Watchdog logs a WARNING when it drops the stale entry; that's

292 # what this test is verifying, so silence it in the test report.

293 caplog.set_level(

@@ -332,8 +328,7 @@ class TestPushScheduler:

328

329 class _StubWriterWorker(NixlPushConnectorWorker):

330 """Construct a worker without invoking ``__init__`` so we can drive

335 - the matching/notif logic without bringing up NIXL or torch.

336 - """

331 + the matching/notif logic without bringing up NIXL or torch."""

332

333 @classmethod

334 def fresh(cls) -> _StubWriterWorker:

@@ -439,8 +434,7 @@ class TestPushWriterRegSend:

434 """``_do_send_reg_notif``: the PUSH_REG goes to every handshaked agent of

435 the P engine, and the agent map is read under ``_handshake_lock`` so the

436 send can't race the handshake done-callback's insert or a concurrent

442 - stale-engine eviction (X1).

443 - """

437 + stale-engine eviction (X1)."""

438

439 def _reg_send_worker(self) -> tuple[_StubWriterWorker, list]:

440 w = _StubWriterWorker.fresh()

@@ -485,8 +479,7 @@ class TestPushWriterRegSend:

479 def test_agent_read_is_serialized_with_handshake_lock(self):

480 """While another thread holds ``_handshake_lock`` (as the handshake

481 done-callback does while writing ``_remote_agents``), the writer's

488 - check-then-send must not run ahead of the critical section.

489 - """

482 + check-then-send must not run ahead of the critical section."""

483 w, failures = self._reg_send_worker()

484 w._remote_agents["prefill-engine"] = {(0, 0): "agent-0"}

485 rd = _registration_data("req-1")

@@ -596,8 +589,7 @@ class TestPushWriterMatching:

589 class TestPushWriterStartLoadKv:

590 def test_finished_blocks_inbox_matches_stashed_registration(self):

591 """Run the writer-loop's finished-blocks drain against a

599 - pre-populated _pending_d_registrations entry.

600 - """

592 + pre-populated _pending_d_registrations entry."""

593 w = _StubWriterWorker.fresh()

594 w._pending_d_registrations["req-C"] = _registration_data("req-C")

595

@@ -622,8 +614,7 @@ class TestPushWriterStartLoadKv:

614

615 def test_start_load_kv_enqueues_to_writer(self):

616 """``start_load_kv`` should hand registrations + finished blocks

625 - to the writer queues without doing matching itself.

626 - """

617 + to the writer queues without doing matching itself."""

618 w = _StubWriterWorker.fresh()

619 # Stub heartbeats to a no-op; tests don't exercise the heartbeat

620 # path here.

@@ -681,8 +672,7 @@ def test_do_start_push_kv_defers_then_writes_when_handshake_ready():

672 NIXL op from the writer or the executor callback); once it resolves the

673 request is re-queued on ``_deferred_push_inbox`` with the wake set; and on

674 re-drive with the handshake ready the WRITE is issued with a correct

684 - ReqMeta.

685 - """

675 + ReqMeta."""

676 w = _StubWriterWorker.fresh()

677 w._logical_to_kernel_block_ids = lambda x, ratio: x

678 xfer_calls: list[dict[str, Any]] = []

@@ -722,8 +712,7 @@ def test_do_start_push_kv_defers_then_writes_when_handshake_ready():

712 def test_do_start_push_kv_drops_request_on_handshake_failure():

713 """Handshake raises: the request is dropped (not re-queued, no WRITE) and

714 the failure is logged. Blocks are reclaimed by the lease/watchdog, matching

725 - the old blocking behaviour.

726 - """

715 + the old blocking behaviour."""

716 w = _StubWriterWorker.fresh()

717 w._logical_to_kernel_block_ids = lambda x, ratio: x

718 xfer_calls: list[dict[str, Any]] = []

@@ -745,8 +734,7 @@ def test_do_start_push_kv_drops_request_on_handshake_failure():

734

735 def test_writer_loop_drains_deferred_push_inbox():

736 """The writer loop drains ``_deferred_push_inbox`` and re-drives

748 - ``_do_start_push_kv`` for each entry (event-driven, no polling).

749 - """

737 + ``_do_start_push_kv`` for each entry (event-driven, no polling)."""

738 w = _StubWriterWorker.fresh()

739 w.nixl_wrapper = MagicMock()

740 w.nixl_wrapper.get_new_notifs.return_value = {}

@@ -779,8 +767,7 @@ def test_writer_loop_drains_deferred_push_inbox():

767

768 def _eviction_worker(engine_ttl: float) -> NixlPushConnectorWorker:

769 """A push worker wired to drive the base ``_ensure_handshake`` eviction

782 - path (``_evict_stale_engines`` + ``_cleanup_remote_engine``).

783 - """

770 + path (``_evict_stale_engines`` + ``_cleanup_remote_engine``)."""

771 w = _StubWriterWorker.fresh()

772 w._engine_ttl = engine_ttl

773 w._engine_last_active = {}

@@ -802,8 +789,7 @@ def _eviction_worker(engine_ttl: float) -> NixlPushConnectorWorker:

789 def test_active_push_refreshes_engine_last_active():

790 """The base ``_ensure_handshake`` refreshes liveness only on a new

791 handshake, so an active push to an already-connected engine must refresh

805 - ``_engine_last_active`` itself or it is reaped mid-stream (S1).

806 - """

792 + ``_engine_last_active`` itself or it is reaped mid-stream (S1)."""

793 w = _eviction_worker(engine_ttl=3600.0)

794 # Already-connected D -> _ensure_handshake returns None, WRITE runs inline.

795 w._remote_agents["decode-engine"] = {(0, 0): "agent-decode"}

@@ -819,8 +805,7 @@ def test_active_push_refreshes_engine_last_active():

805 def test_stale_engine_evicted_on_push():

806 """A push drives ``_ensure_handshake`` -> ``_evict_stale_engines``, so a D

807 engine silent past its TTL is reaped -- no _remote_agents / NIXL agent leak

822 - across D scale up/down (S1).

823 - """

808 + across D scale up/down (S1)."""

809 w = _eviction_worker(engine_ttl=30.0)

810 w._remote_agents["D-old"] = {(0, 0): "agent-D-old"}

811 w._engine_last_active["D-old"] = time.perf_counter() - 10_000.0

@@ -837,8 +822,7 @@ def test_stale_engine_evicted_on_push():

822 def test_cleanup_remote_engine_pops_under_handshake_lock():

823 """Regression (X1): eviction must not remove ``_remote_agents`` entries

824 while another thread is inside the lock's critical section (e.g. the

840 - push writer reading the map to send a registration).

841 - """

825 + push writer reading the map to send a registration)."""

826 w = _eviction_worker(engine_ttl=30.0)

827 w._remote_agents["D-old"] = {(0, 0): "agent-D-old"}

828

@@ -863,8 +847,7 @@ def test_cleanup_remote_engine_pops_under_handshake_lock():

847 class TestPushWriterNotifs:

848 def test_get_new_notifs_processes_forwarded_completion_notif(self):

849 """Non-PUSH_REG notifs forwarded by the writer thread are drained

866 - on the engine main thread inside ``_get_new_notifs``.

867 - """

850 + on the engine main thread inside ``_get_new_notifs``."""

851 w = _StubWriterWorker.fresh()

852 # Pretend the writer thread already forwarded a completion notif

853 # for a request whose KV is being received.

@@ -910,8 +893,7 @@ class TestPushWriterNotifs:

893 @staticmethod

894 def _pollable_worker() -> _StubWriterWorker:

895 """Stub worker with the extra base-worker state the real

913 - ``get_transfer_results`` needs to poll ``_sending_transfers``.

914 - """

896 + ``get_transfer_results`` needs to poll ``_sending_transfers``."""

897 w = _StubWriterWorker.fresh()

898 w.transfer_topo = MagicMock()

899 w.nixl_wrapper = MagicMock()

@@ -953,8 +935,7 @@ class TestPushWriterNotifs:

935 @pytest.mark.parametrize(("pp_size", "is_hma"), [(1, False), (2, True)])

936 def test_completed_push_send_waits_for_consumer_notif(self, pp_size, is_hma):

937 """P-side sends are completed by consumer notifs / lease expiry,

956 - not by local WRITE-handle completion.

957 - """

938 + not by local WRITE-handle completion."""

939 w = self._pollable_worker()

940 w.pp_size = pp_size

941 w._is_hma_required = is_hma

@@ -1025,8 +1006,7 @@ class TestPushSchedulerNegative:

1006

1007 def test_request_finished_unfinished_status_does_not_stage(self):

1008 """If a request is still RUNNING, request_finished must not stash

1028 - blocks for the worker (no push needed).

1029 - """

1009 + blocks for the worker (no push needed)."""

1010 sched = make_nixl_push_scheduler()

1011 _stub_sw_clipping(sched)

1012

@@ -1043,8 +1023,7 @@ class TestPushSchedulerNegative:

1023

1024 def test_request_finished_empty_blocks_does_not_arm_lease(self):

1025 """Empty block-id groups should still complete cleanly without

1046 - arming the lease/finished maps.

1047 - """

1026 + arming the lease/finished maps."""

1027 sched = make_nixl_push_scheduler()

1028 _stub_sw_clipping(sched)

1029

@@ -1059,8 +1038,7 @@ class TestPushSchedulerNegative:

1038

1039 def test_update_connector_output_unknown_request_is_noop(self):

1040 """Idempotent cleanup: clearing a request that was never staged

1062 - must not raise or mutate other state.

1063 - """

1041 + must not raise or mutate other state."""

1042 sched = make_nixl_push_scheduler()

1043 _stub_sw_clipping(sched)

1044

@@ -1093,8 +1071,7 @@ class TestPushWriterNegative:

1071

1072 def test_pop_matching_registration_no_match_when_base_ids_differ(self):

1073 """A registration whose base id (after stripping the random suffix)

1096 - does NOT match the lookup request_id must not be popped.

1097 - """

1074 + does NOT match the lookup request_id must not be popped."""

1075 w = _StubWriterWorker.fresh()

1076 # Two unrelated requests: different base UUIDs, so stripping the

1077 # trailing ``-<8 hex>`` suffix still yields different base ids.

@@ -1110,8 +1087,7 @@ class TestPushWriterNegative:

1087

1088 def test_handle_push_reg_with_non_dict_payload_is_dropped(self, caplog):

1089 """msgpack-encoded non-dict payload (e.g. a list) should be

1113 - dropped without raising.

1114 - """

1090 + dropped without raising."""

1091 caplog.set_level(

1092 logging.CRITICAL,

1093 logger=("vllm.distributed.kv_transfer.kv_connector.v1.nixl.push_worker"),

@@ -1139,8 +1115,7 @@ class TestPushWriterNegative:

1115

1116 def test_handle_push_reg_idempotent_for_same_request_id(self):

1117 """Receiving the same PUSH_REG twice (e.g. P retries after a

1142 - flake) keeps the entry staged exactly once and never fires.

1143 - """

1118 + flake) keeps the entry staged exactly once and never fires."""

1119 w = _StubWriterWorker.fresh()

1120 notif = PUSH_REG_NOTIF_PREFIX + msgspec.msgpack.encode(

1121 _registration_data("req-dup")

@@ -1155,8 +1130,7 @@ class TestPushWriterNegative:

1130 """``get_transfer_results`` must enqueue every completed request

1131 in ``done_sending`` so the writer can drop stale matching state.

1132 Unlike the happy-path test, this verifies the *cardinality*: N

1158 - completed requests -> N evictions, in order.

1159 - """

1133 + completed requests -> N evictions, in order."""

1134 w = _StubWriterWorker.fresh()

1135 with patch.object(

1136 NixlPushConnectorWorker.__mro__[1],

@@ -1179,8 +1153,7 @@ class TestPushWriterNegative:

1153 def test_empty_transfer_results_do_not_enqueue_eviction(self):

1154 """If there's nothing newly done, no eviction should be enqueued.

1155 The wake event IS still set because ``get_finished`` always wakes

1182 - the writer to drain notifs.

1183 - """

1156 + the writer to drain notifs."""

1157 w = _StubWriterWorker.fresh()

1158 with patch.object(

1159 NixlPushConnectorWorker.__mro__[1],

@@ -1221,8 +1194,7 @@ class TestPushWriterNegative:

1194

1195 def test_get_new_notifs_unknown_request_is_logged_and_skipped(self, caplog):

1196 """A completion notif for a request the worker doesn't know

1224 - about should be logged but not crash.

1225 - """

1197 + about should be logged but not crash."""

1198 caplog.set_level(

1199 logging.CRITICAL,

1200 logger=("vllm.distributed.kv_transfer.kv_connector.v1.nixl.push_worker"),

@@ -1254,8 +1226,7 @@ class TestPushWriterNegative:

1226 def test_get_new_notifs_extends_lease_on_heartbeat(self):

1227 """``HB:`` notifs forwarded by the writer thread must extend the

1228 leases of tracked P-side requests on the engine main thread, and

1257 - ignore request IDs that aren't being tracked.

1258 - """

1229 + ignore request IDs that aren't being tracked."""

1230 w = _StubWriterWorker.fresh()

1231 w.transfer_topo = MagicMock()

1232 # _handle_heartbeat reads ``self._lease_extension`` (set in the

@@ -1291,13 +1262,11 @@ class TestPushWriterNegative:

1262

1263 class TestPushPipelineParallel:

1264 """PP-sharded producer: per-stage completion counting, pp_size plumbing,

1294 - and remote-region slicing.

1295 - """

1265 + and remote-region slicing."""

1266

1267 def test_completion_waits_for_one_notif_per_pp_stage(self):

1268 """Each PP stage WRITEs its own layers and sends one notif; D must

1299 - collect pp_size notifs before reporting the recv done.

1300 - """

1269 + collect pp_size notifs before reporting the recv done."""

1270 w = _StubWriterWorker.fresh()

1271 w.transfer_topo = MagicMock()

1272 request_id = "req-pp-2"

@@ -1318,8 +1287,7 @@ class TestPushPipelineParallel:

1287

1288 def test_req_meta_reads_pp_size_from_kv_transfer_params(self):

1289 """D learns the producer's pp_size from kv_transfer_params (forwarded

1321 - by the proxy) and defaults to 1 when absent.

1322 - """

1290 + by the proxy) and defaults to 1 when absent."""

1291 metadata = NixlConnectorMetadata()

1292 params = {

1293 "remote_block_ids": ([0],),

@@ -1342,8 +1310,7 @@ class TestPushPipelineParallel:

1310 this worker holds only a contiguous layer slice; add_remote_agent

1311 trims the remote region list to [offset : offset + num_local_regions]

1312 before building descriptors. We stop right after the slice via a

1345 - sentinel on the next collaborator call.

1346 - """

1313 + sentinel on the next collaborator call."""

1314 block_len = 4096 * 16

1315 w = _StubWriterWorker.fresh() # seeds writer-thread state for teardown

1316 w.pp_size = 2

@@ -1385,8 +1352,7 @@ class TestPushWriterMlaReplication:

1352 """MLA latent KV is replicated across D's TP ranks, so when D_TP > P_TP

1353 (``tp_ratio < 0``) the producer must *WRITE* the latent into every D rank

1354 it handshook -- not write one and merely notify the rest, which would

1388 - leave the un-written ranks decoding against stale KV.

1389 - """

1355 + leave the un-written ranks decoding against stale KV."""

1356

1357 @staticmethod

1358 def _mla_worker_writing_to(d_ranks):

@@ -1611,8 +1577,7 @@ class TestPushPrefixCaching:

1577

1578 def test_partial_prefix_hit_end_trims_producer_blocks(self):

1579 """D registered only its 2 uncomputed suffix blocks; P finished the

1614 - full 5-block sequence. P must WRITE its LAST 2 blocks into D's slots.

1615 - """

1580 + full 5-block sequence. P must WRITE its LAST 2 blocks into D's slots."""

1581 w, _ = self._worker_driving_xfer()

1582 reg = _registration_data("req-pc", local_block_ids=([500, 501],))

1583

tests/v1/kv_connector/unit/test_nixl_rocm_gpu_mem_diag.py

+1 -2

@@ -171,8 +171,7 @@ def test_gpu_memory_nixl_hma(model_name, sw_size):

171 @pytest.mark.parametrize("model_name", ["google/gemma-3-1b-it"])

172 def test_gpu_memory_no_nixl_baseline(model_name):

173 """Same workload without NixlConnector. Comparing driver-level memory

174 - between this and test_gpu_memory_nixl_hma isolates UCX/NIXL impact.

175 - """

174 + between this and test_gpu_memory_nixl_hma isolates UCX/NIXL impact."""

175 from vllm import LLM, SamplingParams

176 from vllm.distributed.parallel_state import cleanup_dist_env_and_memory

177

tests/v1/kv_connector/unit/test_nixl_simple_cpu_offload.py

+3 -6

@@ -213,8 +213,7 @@ def _multi_connector_config(swa_enabled: bool = False):

213 def test_nixl_wins_load_over_cpu_offload():

214 """When NixlConnector (index 0) has matched tokens from a remote prefill, it should

215 win the load: Nixl metadata tracks the recv while CPU offload metadata has no load

216 - scheduled.

217 - """

216 + scheduled."""

217 vllm_config, kv_cache_config = _multi_connector_config()

218 scheduler = create_scheduler(vllm_config, kv_cache_config=kv_cache_config)

219 mc = scheduler.connector

@@ -251,8 +250,7 @@ def test_nixl_wins_load_over_cpu_offload():

250 @patch(NIXL_WRAPPER_PATCH, FakeNixlWrapper)

251 def test_cpu_offload_wins_when_nixl_has_no_match():

252 """When NixlConnector returns 0 matched tokens and SimpleCPUOffloadConnector has a

254 - CPU cache hit, the CPU offload connector (index 1) wins the load.

255 - """

253 + CPU cache hit, the CPU offload connector (index 1) wins the load."""

254 vllm_config, kv_cache_config = _multi_connector_config()

255 scheduler = create_scheduler(vllm_config, kv_cache_config=kv_cache_config)

256 mc = scheduler.connector

@@ -311,8 +309,7 @@ def test_cpu_offload_wins_when_nixl_has_no_match():

309 def test_request_finished_no_async_save(swa_enabled: bool):

310 """A normal request (no P/D) produces no async save from either connector.

311 MultiConnector returns (False, None) via both request_finished and

314 - request_finished_all_groups, and cleans up _requests_to_connector.

315 - """

312 + request_finished_all_groups, and cleans up _requests_to_connector."""

313 from vllm.v1.request import RequestStatus

314

315 vllm_config, kv_cache_config = _multi_connector_config(swa_enabled=swa_enabled)

tests/v1/kv_connector/unit/test_offloading_connector.py

+2 -4

@@ -321,8 +321,7 @@ def test_cpu_offloading(

321 @pytest.mark.skipif(not current_platform.is_cuda(), reason="Requires CUDA")

322 def test_cpu_offloading_metrics() -> None:

323 """Verify that offloading Prometheus metrics (new flat and deprecated

324 - labeled) are emitted after stores and loads.

325 - """

324 + labeled) are emitted after stores and loads."""

325 extra_config: dict = {

326 "cpu_bytes_to_use": 500 << 20,

327 "block_size": CPU_BLOCK_SIZES,

@@ -511,8 +510,7 @@ def test_tiering_offloading() -> None:

510

511 def test_fs_tiering_offloading(tmp_path) -> None:

512 """Tests OffloadingConnector with TieringOffloadingSpec

514 - + fs secondary tier.

515 - """

513 + + fs secondary tier."""

514 extra_config: dict = {

515 "cpu_bytes_to_use": 1 << 30,

516 "block_size": CPU_BLOCK_SIZES,

tests/v1/kv_connector/unit/test_remote_prefill_lifecycle.py

+1 -2

@@ -704,8 +704,7 @@ def test_async_load_reserves_blocks_for_inflight():

704

705 def test_async_loads_both_admitted_when_pool_fits():

706 """Sanity: with a pool large enough, the reservation gate admits both async

707 - loads (it is not over-conservative).

708 - """

707 + loads (it is not over-conservative)."""

708 vllm_config = create_vllm_config()

709 BLOCK_SIZE = vllm_config.cache_config.block_size

710 scheduler = create_scheduler(vllm_config, num_blocks=64)

tests/v1/kv_connector/unit/test_simple_cpu_offload_connector.py

+1 -2

@@ -129,8 +129,7 @@ def _make_ring_kv_cache_config(num_blocks: int = 16) -> KVCacheConfig:

129

130 def test_ring_scratch_group_is_accepted_and_never_offloaded() -> None:

131 """The ring group's capacity is not a token granularity, it holds no

132 - hashed blocks, and it never participates in CPU hits.

133 - """

132 + hashed blocks, and it never participates in CPU hits."""

133 vllm_config = create_vllm_config(

134 kv_connector="SimpleCPUOffloadConnector",

135 kv_role="kv_both",

tests/v1/kv_connector/unit/test_tp_mapping.py

+2 -4

@@ -235,8 +235,7 @@ class TestMambaPlanSplitHandles:

235

236 def test_hetero_block_size_splits(self):

237 """With a block-size ratio, single-source FA sub-block descs pass

238 - through whole; SSM descs are unexpanded and split per source.

239 - """

238 + through whole; SSM descs are unexpanded and split per source."""

239 plan = TPMapping(

240 source_ranks_per_group=((0,), (0, 1)),

241 all_source_ranks=(0, 1),

@@ -271,8 +270,7 @@ class TestMambaPlanSplitHandles:

270

271 def test_hetero_block_size_head_sharded_asserts(self):

272 """Head-sharded FA reads (multiple FA sources) are incompatible with

274 - a block-size mismatch and must fail loudly.

275 - """

273 + a block-size mismatch and must fail loudly."""

274 plan = TPMapping(

275 source_ranks_per_group=((0, 1), (0, 1)),

276 all_source_ranks=(0, 1),

tests/v1/kv_offload/cpu/policies/test_factory.py

+5 -10

@@ -15,8 +15,7 @@ from vllm.v1.kv_offload.cpu.policies.lru import LRUCachePolicy

15 class _DummyCachePolicy(CachePolicy):

16 """Minimal CachePolicy for CachePolicyFactory registration tests. Loaded

17 by module path, so it must be importable at module scope (mirrors

18 - tests/v1/kv_offload/test_factory.py's SingleArgExternalOffloadingSpec).

19 - """

18 + tests/v1/kv_offload/test_factory.py's SingleArgExternalOffloadingSpec)."""

19

20 def __init__(self, cache_capacity: int) -> None:

21 self.cache_capacity = cache_capacity

@@ -55,8 +54,7 @@ class TestCachePolicyFactory:

54

55 def test_pre_registered_policies_can_be_imported(self):

56 """If someone moves a policy module but forgets to update

58 - factory.py, CI fails.

59 - """

57 + factory.py, CI fails."""

58 for name in CachePolicyFactory._registry:

59 cls = CachePolicyFactory._registry[name]()

60 assert issubclass(cls, CachePolicy)

@@ -90,8 +88,7 @@ class TestCachePolicyFactory:

88 register_cache_policy() call -- this is how external projects

89 integrate a custom CachePolicy without forking/patching vLLM.

90 Mirrors tests/v1/kv_offload/test_factory.py's

93 - test_dynamic_load_via_spec_module_path.

94 - """

91 + test_dynamic_load_via_spec_module_path."""

92 policy_cls = CachePolicyFactory.get_cache_policy_cls(

93 "_DummyCachePolicy", "tests.v1.kv_offload.cpu.policies.test_factory"

94 )

@@ -99,8 +96,7 @@ class TestCachePolicyFactory:

96

97 def test_manager_resolves_policy_via_module_path(self):

98 """End-to-end: CPUOffloadingManager resolves an unregistered policy

102 - purely from cache_policy_module_path.

103 - """

99 + purely from cache_policy_module_path."""

100 manager = CPUOffloadingManager(

101 num_chunks=4,

102 cache_policy="_DummyCachePolicy",

@@ -110,7 +106,6 @@ class TestCachePolicyFactory:

106

107 def test_unregistered_policy_without_module_path_raises(self):

108 """eviction_policy not in registry + no cache_policy_module_path ->

113 - ValueError, same shape as the OffloadingSpecFactory error path.

114 - """

109 + ValueError, same shape as the OffloadingSpecFactory error path."""

110 with pytest.raises(ValueError, match="Unknown cache policy"):

111 CachePolicyFactory.get_cache_policy_cls("nonexistent", None)

tests/v1/kv_offload/cpu/test_canonical_layout.py

+3 -6

@@ -126,8 +126,7 @@ def _canonical_handler(gpu_tensor, cpu_tensor, mapping, gpu_to_cpu):

126 def test_gpu_roundtrip_assembles_canonical_page_across_ranks():

127 """Two TP2-style rank handlers must scatter into one shared canonical

128 CPU page such that each rank reloads bit-exact and a whole-page (TP1)

129 - reader sees both shards — the cross-topology contract.

130 - """

129 + reader sees both shards — the cross-topology contract."""

130 torch.manual_seed(0)

131 num_blocks = 4

132 gpu_rank = [

@@ -204,8 +203,7 @@ def _nhd_shard_mapping(tp: int, rank: int) -> CanonicalPageMapping:

203

204 def _head_shard(full_kv: torch.Tensor, tp: int, rank: int) -> torch.Tensor:

205 """This rank's local page rows out of the (blocks, 2, tokens, heads,

207 - head_bytes) ground truth.

208 - """

206 + head_bytes) ground truth."""

207 local_heads = _TOTAL_HEADS // tp

208 shard = full_kv[:, :, :, rank * local_heads : (rank + 1) * local_heads, :]

209 return shard.reshape(full_kv.shape[0], -1).contiguous()

@@ -216,8 +214,7 @@ def _head_shard(full_kv: torch.Tensor, tp: int, rank: int) -> torch.Tensor:

214 def test_cross_topology_roundtrip(writer_tp: int, reader_tp: int):

215 """KV written at one tp must be readable at another: writer ranks scatter

216 head shards into a shared canonical region, reader ranks gather their own

219 - shards, and every reader must see the writers' ground-truth bytes.

220 - """

217 + shards, and every reader must see the writers' ground-truth bytes."""

218 torch.manual_seed(0)

219 num_blocks = 3

220 full_kv = torch.randint(

tests/v1/kv_offload/cpu/test_shared_offload_region.py

+14 -28

@@ -167,8 +167,7 @@ def _mp_race_construct_and_write(

167 ) -> None:

168 """Race to construct a SharedOffloadRegion, write fill_value, then wait

169 for the parent's cleanup signal before tearing down. The wait gives the

170 - parent a window to read the raw mmap before the creator removes the file.

171 - """

170 + parent a window to read the raw mmap before the creator removes the file."""

171 try:

172 region = SharedOffloadRegion(

173 engine_id=engine_id,

@@ -197,8 +196,7 @@ def _mp_barrier_construct_and_hold(

196 replicated: bool = False,

197 ) -> None:

198 """Construct with a real cross-process barrier, write, then hold the

200 - mapping (no cleanup) until the parent SIGKILLs this process.

201 - """

199 + mapping (no cleanup) until the parent SIGKILLs this process."""

200 try:

201 populate_calls = 0

202 get_populate_write_fn = region_module._get_populate_write_fn

@@ -252,8 +250,7 @@ def iid():

250

251 def test_create_next_worker_view_shape_and_stride(iid):

252 """Returned tensor must have shape (num_chunks, tensor_page_size) and

255 - stride (row_stride, 1) where row_stride = cpu_page_size * num_workers.

256 - """

253 + stride (row_stride, 1) where row_stride = cpu_page_size * num_workers."""

254 with _region(iid, num_chunks=4, cpu_page_size=2 * PAGE_SIZE) as r:

255 t = r.create_next_worker_view(PAGE_SIZE)

256 assert t.shape == (4, PAGE_SIZE)

@@ -293,8 +290,7 @@ def test_create_next_worker_view_row_stride_with_multiple_workers(iid):

290

291 def test_create_next_worker_view_cursor_advances(iid):

292 """Each create_next_worker_view call must advance _worker_offset by

296 - tensor_page_size.

297 - """

293 + tensor_page_size."""

294 with _region(iid, cpu_page_size=3 * PAGE_SIZE) as r:

295 assert r._worker_offset == 0

296 r.create_next_worker_view(PAGE_SIZE)

@@ -351,8 +347,7 @@ def test_create_next_worker_view_overflow_does_not_mutate_cursor(iid):

347

348 def test_create_next_worker_view_write_visible_in_raw_mmap(iid):

349 """Writes into a create_next_worker_view view must appear at the correct

354 - raw mmap offset

355 - """

350 + raw mmap offset"""

351 with _region(iid, num_chunks=4) as r:

352 t = r.create_next_worker_view(PAGE_SIZE)

353 t[2, :] = 42 # write to chunk row 2

@@ -385,8 +380,7 @@ def test_create_next_worker_view_multi_tensor_layout(iid):

380

381 def test_create_next_worker_view_multiprocess_slots(iid):

382 """Each worker process calls create_next_worker_view and writes distinct data;

388 - the parent verifies each slot lands at the correct interleaved offset.

389 - """

383 + the parent verifies each slot lands at the correct interleaved offset."""

384 num_workers = 2

385 num_chunks = 4

386

@@ -671,8 +665,7 @@ def test_multi_worker_race_shared_memory_visible(iid):

665

666 def test_multiprocess_race_construct_and_write(iid):

667 """N processes race to construct the same SharedOffloadRegion, each writes

674 - fill_value = rank+1 into their slot; parent verifies interleaved layout.

675 - """

668 + fill_value = rank+1 into their slot; parent verifies interleaved layout."""

669 num_workers = 4

670 num_chunks = 3

671

@@ -798,8 +791,7 @@ def test_cleanup_unregisters_every_pinned_chunk(iid, monkeypatch):

791 def test_cleanup_after_create_next_worker_view_releases_mmap(iid):

792 """cleanup() must close the mmap even after create_next_worker_view was called.

793 create_next_worker_view returns a view that shares storage with _base; both must be

801 - released before mmap.close() can succeed.

802 - """

794 + released before mmap.close() can succeed."""

795 r = _make_region(iid)

796 mmap_obj = r.mmap_obj

797

@@ -978,8 +970,7 @@ def test_ftruncate_failure_cleans_up_creator(monkeypatch):

970

971 def test_backing_file_unlinked_after_barrier(iid):

972 """The file must exist until the barrier releases (late joiners need the

981 - name) and be gone right after, so no exit path can leak it.

982 - """

973 + name) and be gone right after, so no exit path can leak it."""

974 path = f"/dev/shm/vllm_offload_{iid}.mmap"

975 seen_at_barrier = []

976

@@ -1001,8 +992,7 @@ def test_backing_file_unlinked_after_barrier(iid):

992

993 def test_barrier_failure_unlinks_creator_and_raises(iid):

994 """A failed rendezvous must remove the creator's file and re-raise, not

1004 - leave a stub that wedges the next start in _wait_for_file_size.

1005 - """

995 + leave a stub that wedges the next start in _wait_for_file_size."""

996 path = f"/dev/shm/vllm_offload_{iid}.mmap"

997 with pytest.raises(RuntimeError, match="peer died"):

998 _make_region(iid, barrier=MagicMock(side_effect=RuntimeError("peer died")))

@@ -1012,8 +1002,7 @@ def test_barrier_failure_unlinks_creator_and_raises(iid):

1002 @pytest.mark.parametrize("replicated", [False, True])

1003 def test_mp_barrier_unlinks_file_and_survives_sigkill(iid, replicated):

1004 """With a real cross-process barrier every worker maps one shared inode,

1015 - the name is gone while they run, and SIGKILL leaks nothing.

1016 - """

1005 + the name is gone while they run, and SIGKILL leaks nothing."""

1006 num_workers = 2

1007 path = f"/dev/shm/vllm_offload_{iid}.mmap"

1008 ctx = get_mp_context()

@@ -1052,8 +1041,7 @@ def test_mp_barrier_unlinks_file_and_survives_sigkill(iid, replicated):

1041 @pytest.mark.parametrize("failure_phase", ["shm", "memory", "population"])

1042 def test_setup_failure_before_barrier_releases_peers(iid, monkeypatch, failure_phase):

1043 """A worker that dies before the rendezvous must still arrive at the

1055 - barrier, or its peers block in the collective until it times out.

1056 - """

1044 + barrier, or its peers block in the collective until it times out."""

1045 failure = MagicMock(side_effect=ValueError("Insufficient space"))

1046 memory_check = failure if failure_phase == "memory" else None

1047 if failure_phase == "shm":

@@ -1084,8 +1072,7 @@ def test_setup_failure_before_barrier_releases_peers(iid, monkeypatch, failure_p

1072

1073 def test_mmap_failure_unlinks_creator_before_releasing_peers(iid, monkeypatch):

1074 """A creator that fails after sizing the file must drop it before arriving

1087 - at the barrier, so the next start does not land on a stale file.

1088 - """

1075 + at the barrier, so the next start does not land on a stale file."""

1076 path = f"/dev/shm/vllm_offload_{iid}.mmap"

1077 monkeypatch.setattr("mmap.mmap", MagicMock(side_effect=OSError("mmap")))

1078 seen_at_barrier = []

@@ -1098,8 +1085,7 @@ def test_mmap_failure_unlinks_creator_before_releasing_peers(iid, monkeypatch):

1085

1086 def test_barrier_release_failure_keeps_original_error(iid, monkeypatch):

1087 """When releasing the peers fails too, the setup error that explains the

1101 - failure must be the one that propagates.

1102 - """

1088 + failure must be the one that propagates."""

1089 import vllm.v1.kv_offload.cpu.shared_offload_region as region

1090

1091 monkeypatch.setattr(

tests/v1/kv_offload/tiering/p2p/test_manager.py

+31 -62

@@ -48,8 +48,7 @@ def _remote_prefiller_kv_params(

48 kv_request_id: str = "req-1",

49 ) -> dict:

50 """Decoder-side kv_transfer_params: ``remote_prefiller`` sub-dict carries

51 - kv_request_id + remote_host + remote_port.

52 - """

51 + kv_request_id + remote_host + remote_port."""

52 return {

53 "remote_prefiller": {

54 "kv_request_id": kv_request_id,

@@ -65,8 +64,7 @@ def _remote_kv_source_kv_params(

64 kv_request_id: str = "req-1",

65 ) -> dict:

66 """Symmetric-P2P consumer kv_transfer_params: ``remote_kv_source`` sub-dict has

68 - the same shape as ``remote_prefiller`` (kv_request_id + remote_host + port).

69 - """

67 + the same shape as ``remote_prefiller`` (kv_request_id + remote_host + port)."""

68 return {

69 "remote_kv_source": {

70 "kv_request_id": kv_request_id,

@@ -78,8 +76,7 @@ def _remote_kv_source_kv_params(

76

77 def _remote_decoder_kv_params(kv_request_id: str = "req-1") -> dict:

78 """Prefiller-side kv_transfer_params: ``remote_decoder`` sub-dict carries

81 - kv_request_id only.

82 - """

79 + kv_request_id only."""

80 return {"remote_decoder": {"kv_request_id": kv_request_id}}

81

82

@@ -244,8 +241,7 @@ class TestLookup:

241 def test_lookup_returns_miss_without_prefill_key(self):

242 """No ``remote_prefiller`` sub-dict means the request was not routed for

243 remote prefill — local prefill should run instead, so lookup()

247 - returns MISS even when a stale ``remote_decoder`` block is present.

248 - """

244 + returns MISS even when a stale ``remote_decoder`` block is present."""

245 mgr = _make_manager()

246 ctx = _req_context(kv_params=_remote_decoder_kv_params())

247 assert mgr.lookup(b"key", ctx) is LookupResult.MISS

@@ -313,8 +309,7 @@ class TestServeExternalRequests:

309 def test_flushes_failed_serve_ctxs_then_serves_each_session(self):

310 """serve_external_requests releases the failed serves left by

311 reaped sessions via parent.on_request_finished (clearing the

316 - queue), then delegates to every live session with the same parent.

317 - """

312 + queue), then delegates to every live session with the same parent."""

313 mgr = _make_manager()

314 ctx = ReqContext(req_id="p2p:peer:req-1:lu1")

315 mgr._failed_serve_ctxs = [ctx]

@@ -368,8 +363,7 @@ class TestSubmitStore:

363 def test_no_binding_yet_parks_in_unbound_stores(self):

364 """submit_store without a bound session buffers the batch keyed

365 by kv_request_id; no session is pre-created (peer_id is unknown

371 - to the producer at store time).

372 - """

366 + to the producer at store time)."""

367 mgr = _make_manager()

368 job = _job_metadata(

369 job_id=1,

@@ -396,8 +390,7 @@ class TestSubmitStore:

390 def test_routes_to_bound_session(self):

391 """If a session has already received FetchMsg for this

392 kv_request_id (so _kv_to_session is populated), submit_store

399 - forwards directly to that session rather than re-buffering.

400 - """

393 + forwards directly to that session rather than re-buffering."""

394 mgr = _make_manager()

395 bound = _FakeSession(peer_id="10.0.0.1:8000", connected=True)

396 mgr._kv_to_session["req-1"] = bound # type: ignore[assignment]

@@ -420,8 +413,7 @@ class TestSubmitStore:

413 def test_extra_top_level_keys_are_ignored(self):

414 """Producer-side kv_transfer_params should not pre-create a

415 session even when a stale caller still passes a top-level

423 - ``remote_host``/``remote_port`` next to ``remote_decoder``.

424 - """

416 + ``remote_host``/``remote_port`` next to ``remote_decoder``."""

417 mgr = _make_manager()

418 params = _remote_decoder_kv_params()

419 params["remote_host"] = "stale"

@@ -465,8 +457,7 @@ class TestSubmitLoad:

457

458 def test_happy_path_with_active_session(self):

459 """When the peer's session exists, submit_load forwards to

468 - session.request_blocks and does NOT add a finished result yet.

469 - """

460 + session.request_blocks and does NOT add a finished result yet."""

461 mgr = _make_manager()

462 peer_id = "10.0.0.1:8000"

463 existing = _FakeSession(peer_id=peer_id, connected=True)

@@ -486,8 +477,7 @@ class TestSubmitLoad:

477 def test_missing_consumer_flag_fails(self):

478 """Peer fields present but neither do_remote_prefill nor

479 do_p2p_fetch is set — submit_load fails the job rather than

489 - emit a stray FetchMsg.

490 - """

480 + emit a stray FetchMsg."""

481 mgr = _make_manager()

482 params = {

483 "remote_host": "10.0.0.1",

@@ -531,8 +521,7 @@ class TestOnRequestFinished:

521 def test_decoder_side_calls_session_finish_request(self):

522 """Decoder-side finish (``remote_prefiller`` set) still routes via peer_id

523 because the consumer addresses the producer it loaded from. The

534 - session's finish_request cancels the client-role load.

535 - """

524 + session's finish_request cancels the client-role load."""

525 mgr = _make_manager()

526 peer_id = "10.0.0.1:8000"

527 session = _FakeSession(peer_id=peer_id)

@@ -544,8 +533,7 @@ class TestOnRequestFinished:

533 def test_p2p_consumer_side_calls_session_finish_request(self):

534 """Symmetric-P2P consumer finish (``remote_kv_source`` set) routes via peer_id

535 so the session drops any pending lookups (cancel_lookups) and

547 - cancels any inbound load.

548 - """

536 + cancels any inbound load."""

537 mgr = _make_manager()

538 peer_id = "10.0.0.1:8000"

539 session = _FakeSession(peer_id=peer_id)

@@ -556,8 +544,7 @@ class TestOnRequestFinished:

544

545 def test_prefiller_bound_id_routes_via_kv_to_session(self):

546 """Prefiller-side finish for an id whose session is already bound

559 - (FetchMsg received) routes via _kv_to_session and pops the entry.

560 - """

547 + (FetchMsg received) routes via _kv_to_session and pops the entry."""

548 mgr = _make_manager()

549 bound = _FakeSession(peer_id="some-peer:1", connected=True)

550 mgr._kv_to_session["req-1"] = bound # type: ignore[assignment]

@@ -571,8 +558,7 @@ class TestOnRequestFinished:

558 and no session binding is a no-op on `_unbound_stores`. The

559 parked batches survive until a peer fetches them or the

560 `_reap_unbound_stores` timeout fires — `on_request_finished`

574 - must not evict them.

575 - """

561 + must not evict them."""

562 from vllm.v1.kv_offload.tiering.p2p.manager import _UnboundStoreBatch

563

564 mgr = _make_manager()

@@ -742,8 +728,7 @@ class TestGetFinished:

728

729 def test_reap_fails_probes(self):

730 """A reaped session's in-flight lookups land in _failed_req_ids so

745 - the consumer's lookup() returns MISS instead of RETRY forever.

746 - """

731 + the consumer's lookup() returns MISS instead of RETRY forever."""

732

733 class FakeData:

734 def remove_remote_peer(self, pid):

@@ -778,8 +763,7 @@ class TestGetFinished:

763 def test_unbound_store_reaped_after_timeout(self):

764 """Unbound stores past the reap deadline surface as failed and their

765 kv_request_id lands in _reaped_stores so a late FetchMsg is rejected

781 - instead of parking demand nothing can satisfy.

782 - """

766 + instead of parking demand nothing can satisfy."""

767 from vllm.v1.kv_offload.tiering.p2p.manager import _UnboundStoreBatch

768

769 mgr = self._make()

@@ -801,8 +785,7 @@ class TestGetFinished:

785

786 def test_submit_store_parks_unbound_batch(self):

787 """submit_store on an unbound id appends a batch with a fresh

804 - submitted_at stamp so the unbound-store sweep can age it out.

805 - """

788 + submitted_at stamp so the unbound-store sweep can age it out."""

789 mgr = _make_manager()

790 job = _job_metadata(

791 job_id=1, kv_params=_remote_decoder_kv_params(kv_request_id="req-1")

@@ -824,21 +807,18 @@ class TestHasPendingWork:

807 """has_pending_work() must always return True so the engine keeps

808 ticking the offload pipeline — that's the only thread driving

809 _control.poll() (incoming peer connects) and session.poll()

827 - (incoming fetch messages on existing sessions).

828 - """

810 + (incoming fetch messages on existing sessions)."""

811

812 def test_returns_true_unconditionally_to_keep_engine_ticking(self):

813 """Even with no sessions and no jobs, has_pending_work() returns

814 True so the engine keeps calling get_finished_jobs(), which is

833 - what drives _control.poll() for inbound peer connects.

834 - """

815 + what drives _control.poll() for inbound peer connects."""

816 mgr = _make_manager()

817 assert mgr.has_pending_work() is True

818

819 def test_returns_true_even_when_sessions_present(self):

820 """The result is the same regardless of session state — there is

840 - no 'idle' branch.

841 - """

821 + no 'idle' branch."""

822 mgr = _make_manager()

823 mgr._sessions["peer:1"] = _FakeSession(peer_id="peer:1") # type: ignore[assignment]

824 assert mgr.has_pending_work() is True

@@ -851,8 +831,7 @@ class TestHasPendingWork:

831

832 class _ShutdownFakeData:

833 """Fake DataTransport that records cancel/poll/close calls and

854 - drives the wait-cancel loop with a scriptable `still` queue.

855 - """

834 + drives the wait-cancel loop with a scriptable `still` queue."""

835

836 def __init__(self, still_queue: list[list[int]] | None = None) -> None:

837 # Each list in still_queue is the set of ids the next

@@ -896,8 +875,7 @@ class _ShutdownFakeControl:

875 class TestShutdownDrain:

876 """shutdown() drains inflight transfers via cancel(mode='wait')

877 before calling _data.close(), with a 3s deadline fallback to

899 - cancel(mode='immediate').

900 - """

878 + cancel(mode='immediate')."""

879

880 def _prep(

881 self,

@@ -1316,8 +1294,7 @@ class _RecordingConn:

1294

1295 class TestAcceptNewPeers:

1296 """A second inbound from an already-connected peer is rejected and the

1319 - new conn is closed; the existing session is left untouched.

1320 - """

1297 + new conn is closed; the existing session is left untouched."""

1298

1299 def test_duplicate_connection_is_closed_and_existing_session_untouched(self):

1300 mgr = _make_manager()

@@ -1339,8 +1316,7 @@ class TestAcceptNewPeers:

1316 """An inbound conn from a peer with no existing session creates

1317 a fresh connected session and registers it under conn.peer_id.

1318 The prefiller has no pre-created pending session anymore — the

1342 - first signal of a peer's existence is its inbound connection.

1343 - """

1319 + first signal of a peer's existence is its inbound connection."""

1320

1321 class FakeData:

1322 block_len = 4096

@@ -1387,8 +1363,7 @@ class TestAcceptNewPeers:

1363

1364 class TestPollOnce:

1365 """_poll_once must drain control, accept new peers, poll every session,

1390 - surface results, and reap dead sessions — in that order.

1391 - """

1366 + surface results, and reap dead sessions — in that order."""

1367

1368 def test_orchestrates_accept_poll_and_reap(self):

1369 mgr = _make_manager()

@@ -1444,8 +1419,7 @@ class TestPollOnce:

1419 def test_new_fetch_id_binds_and_replays_unbound_batches(self):

1420 """When session.poll() reports a kv_request_id whose FetchMsg

1421 arrived this tick, the manager binds it to that session and

1447 - replays every parked submit_store batch via add_stored_blocks.

1448 - """

1422 + replays every parked submit_store batch via add_stored_blocks."""

1423 from vllm.v1.kv_offload.tiering.p2p.manager import _UnboundStoreBatch

1424

1425 mgr = _make_manager()

@@ -1479,8 +1453,7 @@ class TestPollOnce:

1453

1454 def test_new_fetch_id_with_no_unbound_still_binds(self):

1455 """A FetchMsg for a kv_request_id with no parked batches still

1482 - records the binding so subsequent submit_stores route fast.

1483 - """

1456 + records the binding so subsequent submit_stores route fast."""

1457 mgr = _make_manager()

1458 peer = "10.0.0.1:8000"

1459 sess = _FakeSession(

@@ -1503,8 +1476,7 @@ class TestPollOnce:

1476

1477 def test_failed_load_records_kv_request_id(self):

1478 """A LoadResult(success=False) from session.poll() must add its

1506 - kv_request_id to _failed_req_ids so future lookups return MISS.

1507 - """

1479 + kv_request_id to _failed_req_ids so future lookups return MISS."""

1480 mgr = _make_manager()

1481 peer = "10.0.0.1:8000"

1482 sess = _FakeSession(

@@ -1568,8 +1540,7 @@ class TestDrainJobs:

1540

1541 def test_logs_warning_after_5s_then_completes(self, monkeypatch):

1542 """A session that stays inflight past 5s triggers the warning, and

1571 - once it clears the loop returns.

1572 - """

1543 + once it clears the loop returns."""

1544 mgr = _make_manager()

1545 mgr._control = _DrainCtrl() # type: ignore[assignment]

1546 sess = _FakeSession(peer_id="peer:1")

@@ -1627,8 +1598,7 @@ class TestDrainJobs:

1598 class TestOnScheduleEnd:

1599 def test_is_noop(self):

1600 """on_schedule_end is a documented no-op; just confirm it doesn't

1630 - raise and doesn't mutate state.

1631 - """

1601 + raise and doesn't mutate state."""

1602 mgr = _make_manager()

1603 before_sessions = dict(mgr._sessions)

1604 before_jobs = list(mgr._finished_jobs)

@@ -1654,8 +1624,7 @@ class TestConnectionDeathMidTransfer:

1624 prefiller-side store no longer travels through the session at store

1625 time (it's parked in _unbound_stores keyed by kv_request_id), so its

1626 cleanup on connection death is via on_request_finished or the

1657 - unbound-store timeout — covered separately below.

1658 - """

1627 + unbound-store timeout — covered separately below."""

1628

1629 def test_dead_connection_with_pending_work_surfaces_failures(self):

1630 mgr_a, mgr_b = _build_paired_managers()

tests/v1/kv_offload/tiering/p2p/test_sessions.py

+40 -80

@@ -626,8 +626,7 @@ class TestClientFlows:

626 """collect_results / has_active_loads use the _active_loads work-list,

627 armed when a fetch is issued and discarded exactly when its load

628 clears — a probe-only request never enters it, and completion empties

629 - it while the entry may briefly linger for GC.

630 - """

629 + it while the entry may briefly linger for GC."""

630 session, conn, _ = _make_session()

631 _activate(session, conn)

632 client = session._client

@@ -704,8 +703,7 @@ class TestClientFlows:

703 def test_load_abort_ack_clears_request(self):

704 """After load timeout sends AbortFetch, an arriving AbortAckMsg from

705 the peer surfaces the failure cleanly and removes the request from

707 - _requests — covers the on_abort_ack arrival path.

708 - """

706 + _requests — covers the on_abort_ack arrival path."""

707 session, conn, _ = _make_session()

708 _activate(session, conn)

709 session.request_blocks(

@@ -742,8 +740,7 @@ class TestLookupFlow:

740 def test_aggregate_flush_resolve_round_trip(self):

741 """register_lookup → flush sends one LookupMsg → response

742 resolves entries → register_lookup returns the cached bool on

745 - every call, and repeat probes never re-issue a LookupMsg.

746 - """

743 + every call, and repeat probes never re-issue a LookupMsg."""

744 session, conn, _ = _make_session()

745 _activate(session, conn)

746

@@ -796,8 +793,7 @@ class TestLookupFlow:

793 def test_request_blocks_clears_probe_cache(self):

794 """A resolved HIT probe is popped when its fetch is issued, so a

795 re-scheduled request re-probes instead of trusting the stale True

799 - (the served block is unpinned and may have been evicted).

800 - """

796 + (the served block is unpinned and may have been evicted)."""

797 session, conn, _ = _make_session()

798 _activate(session, conn)

799

@@ -883,8 +879,7 @@ class TestLookupFlow:

879 def test_multiple_lookup_msgs_across_steps(self):

880 """A request's block set may be discovered across scheduler steps:

881 each step that registers new keys flushes its own LookupMsg for

886 - the same kv_request_id, carrying only the newly-probed keys.

887 - """

882 + the same kv_request_id, carrying only the newly-probed keys."""

883 session, conn, _ = _make_session()

884 _activate(session, conn)

885

@@ -912,8 +907,7 @@ class TestLookupFlow:

907

908 def test_split_response_resolves_across_messages(self):

909 """Producer may answer one LookupMsg's keys across multiple

915 - LookupRespMsgs — pairs are self-describing so each lands.

916 - """

910 + LookupRespMsgs — pairs are self-describing so each lands."""

911 session, conn, _ = _make_session()

912 _activate(session, conn)

913

@@ -960,8 +954,7 @@ class TestLookupFlow:

954 def test_finish_after_flushed_lookup_sends_empty_fetch(self):

955 """LookupMsg flushed but no FetchMsg sent (all-miss case) →

956 finish_request emits an empty FetchMsg so the peer can drop its

963 - lookup state and call parent.on_request_finished.

964 - """

957 + lookup state and call parent.on_request_finished."""

958 session, conn, _ = _make_session()

959 _activate(session, conn)

960

@@ -979,8 +972,7 @@ class TestLookupFlow:

972

973 def test_finish_without_flushed_lookup_sends_no_fetch(self):

974 """No LookupMsg was ever sent → finish_request must not emit an

982 - empty FetchMsg (the peer has no state to release).

983 - """

975 + empty FetchMsg (the peer has no state to release)."""

976 session, conn, _ = _make_session()

977 _activate(session, conn)

978

@@ -995,8 +987,7 @@ class TestLookupFlow:

987

988 def test_finish_after_real_fetch_sends_no_second_fetch(self):

989 """A real FetchMsg was already sent for the id → finish_request

998 - must not emit a second (empty) FetchMsg.

999 - """

990 + must not emit a second (empty) FetchMsg."""

991 session, conn, _ = _make_session()

992 _activate(session, conn)

993

@@ -1027,8 +1018,7 @@ class TestLookupFlow:

1018 """``poll()`` only enqueues an inbound LookupMsg — no response is

1019 sent until ``serve_external_requests``. With an all-miss parent

1020 the aggregated LookupRespMsg carries the same keys and

1030 - ``hits=[False, ...]``.

1031 - """

1021 + ``hits=[False, ...]``."""

1022 session, conn, _ = _make_session()

1023 _activate(session, conn)

1024

@@ -1088,8 +1078,7 @@ class TestServerLookupHandling:

1078 def test_immediate_hits_create_one_store_job(self):

1079 """All-HIT batch: one create_store_job call with all keys, one

1080 LookupRespMsg with hits=[True]*N, on_request_finished fires at the

1091 - end of serve, and `available` is populated for the eventual fetch.

1092 - """

1081 + end of serve, and `available` is populated for the eventual fetch."""

1082 cb = FakeParent(stored={b"hA": 1, b"hB": 2, b"hC": 3})

1083 session, conn, _ = _make_session()

1084 _activate(session, conn)

@@ -1119,8 +1108,7 @@ class TestServerLookupHandling:

1108

1109 def test_all_misses_no_store_job_finish_fires(self):

1110 """All-MISS batch: no create_store_job call; one LookupRespMsg

1122 - with hits=[False]*N; on_request_finished fires at end of serve.

1123 - """

1111 + with hits=[False]*N; on_request_finished fires at end of serve."""

1112 cb = FakeParent()

1113 session, conn, _ = _make_session()

1114 _activate(session, conn)

@@ -1140,8 +1128,7 @@ class TestServerLookupHandling:

1128 """HIT/MISS resolutions do not go out on first sight when any

1129 key is still HIT_PENDING / RETRY. The lookup parks until every

1130 key has settled (or the deadline fires), then one

1143 - LookupRespMsg carries all keys in wire order.

1144 - """

1131 + LookupRespMsg carries all keys in wire order."""

1132 cb = FakeParent(

1133 stored={b"hA": 1},

1134 pending={b"hB"},

@@ -1170,8 +1157,7 @@ class TestServerLookupHandling:

1157 the deferred aggregate response: one LookupRespMsg carrying

1158 both keys in wire order, and one create_store_job call per

1159 HIT (the second HIT is pinned when it resolves, not when the

1173 - response goes out).

1174 - """

1160 + response goes out)."""

1161 cb = FakeParent(stored={b"hA": 1}, pending={b"hB"})

1162 session, conn, _ = _make_session()

1163 _activate(session, conn)

@@ -1207,8 +1193,7 @@ class TestServerLookupHandling:

1193 def test_pending_timeout_replies_miss_no_store_job(self):

1194 """A HIT_PENDING key that stays pending past the batch

1195 ``deadline`` is force-MISS and never pinned; the deferred

1210 - aggregate response fires with hits=[False].

1211 - """

1196 + aggregate response fires with hits=[False]."""

1197 cb = FakeParent(pending={b"hA"})

1198 session, conn, _ = _make_session()

1199 _activate(session, conn)

@@ -1235,8 +1220,7 @@ class TestServerLookupHandling:

1220

1221 def test_finish_request_called_per_lookup_msg_not_per_kv_request_id(self):

1222 """Two LookupMsgs for the same kv_request_id get distinct ctxs

1238 - and two on_request_finished calls (one per batch).

1239 - """

1223 + and two on_request_finished calls (one per batch)."""

1224 cb = FakeParent(stored={b"hA": 1, b"hB": 2})

1225 session, conn, _ = _make_session()

1226 _activate(session, conn)

@@ -1260,8 +1244,7 @@ class TestServerLookupHandling:

1244 """Tearing the session down with a parked batch returns the

1245 synthetic ctx as a failed serve (no parent handle at teardown) so

1246 the manager can release the TieringManager's state on its next

1263 - serve.

1264 - """

1247 + serve."""

1248 cb = FakeParent(pending={b"hA"})

1249 session, conn, _ = _make_session()

1250 _activate(session, conn)

@@ -1282,8 +1265,7 @@ class TestServerLookupHandling:

1265 def test_wire_finish_drops_pending_batches_for_kv_request_id(self):

1266 """``ServerRole.finish(kv_request_id)`` drops every parked batch

1267 whose kv_request_id matches and queues its ctx for the next

1285 - serve's on_request_finished.

1286 - """

1268 + serve's on_request_finished."""

1269 cb = FakeParent(pending={b"hA", b"hB"})

1270 session, conn, _ = _make_session()

1271 _activate(session, conn)

@@ -1314,8 +1296,7 @@ class TestServerLookupHandling:

1296 def test_incoming_fetch_drops_pending_lookups_for_kv_request_id(self):

1297 """A peer FetchMsg terminates the lookup phase for its id: parked

1298 lookups with matching kv_request_id are dropped and their

1317 - ctx queued for on_request_finished; other kv_request_ids untouched.

1318 - """

1299 + ctx queued for on_request_finished; other kv_request_ids untouched."""

1300 cb = FakeParent(pending={b"hA", b"hB"})

1301 session, conn, _ = _make_session()

1302 _activate(session, conn)

@@ -1357,8 +1338,7 @@ class TestServerLookupHandling:

1338 def test_lookup_then_fetch_round_trip_emits_store_result(self):

1339 """End-to-end: lookup pins primary slots → fetch matches them →

1340 NIXL transfer completes → StoreResult surfaces with the

1360 - create_store_job's job_id (the engine releases the pin).

1361 - """

1341 + create_store_job's job_id (the engine releases the pin)."""

1342 cb = FakeParent(stored={b"hA": 7, b"hB": 8})

1343 session, conn, transport = _make_session()

1344 _activate(session, conn)

@@ -1480,8 +1460,7 @@ class TestServerFlows:

1460

1461 def test_abort_fetch_defers_ack_when_cancel_pending(self):

1462 """If cancel(mode='wait') reports still-inflight tids, the ack is

1483 - deferred and the abort is parked (abort_started_at set).

1484 - """

1463 + deferred and the abort is parked (abort_started_at set)."""

1464 session, conn, transport = _make_session()

1465 _activate(session, conn)

1466 # Seed an inflight transfer for req-1 that the transport pretends

@@ -1516,8 +1495,7 @@ class TestServerFlows:

1495

1496 def test_abort_fetch_acks_after_drain(self):

1497 """Once the transport reports the tid as DONE the parked abort

1519 - completes and AbortAckMsg is sent.

1520 - """

1498 + completes and AbortAckMsg is sent."""

1499 session, conn, transport = _make_session()

1500 _activate(session, conn)

1501 tid = 42

@@ -1556,8 +1534,7 @@ class TestServerFlows:

1534

1535 def test_abort_fetch_force_cancels_after_timeout(self):

1536 """If wait-mode never drains, the deadline forces immediate

1559 - cancel and an ack is still sent.

1560 - """

1537 + cancel and an ack is still sent."""

1538 session, conn, transport = _make_session()

1539 _activate(session, conn)

1540 tid = 42

@@ -1599,8 +1576,7 @@ class TestServerFlows:

1576

1577 def test_abort_fetch_idempotent_while_draining(self):

1578 """Receiving AbortFetchMsg twice for the same kv_request_id

1602 - keeps a single pending entry and produces a single ack.

1603 - """

1579 + keeps a single pending entry and produces a single ack."""

1580 session, conn, transport = _make_session()

1581 _activate(session, conn)

1582 tid = 42

@@ -1658,8 +1634,7 @@ class TestServerFlows:

1634 def test_store_timeout_then_late_completion_no_duplicate(self):

1635 """A job timed out by _timeout_pending_store_jobs must not also

1636 emit a contradictory StoreResult(success=True) when the transport

1661 - later reports the same transfer as done.

1662 - """

1637 + later reports the same transfer as done."""

1638 session, conn, transport = _make_session()

1639 _activate(session, conn)

1640 session.add_stored_blocks("req-1", [b"k1"], [0], job_id=1)

@@ -1692,8 +1667,7 @@ class TestServerFlows:

1667 def test_store_timeout_then_late_failure_no_duplicate(self):

1668 """Symmetric guard: a timed-out job must not also emit a second

1669 StoreResult(success=False) when the transport later reports the

1695 - same transfer as failed.

1696 - """

1670 + same transfer as failed."""

1671 session, conn, transport = _make_session()

1672 _activate(session, conn)

1673 session.add_stored_blocks("req-1", [b"k1"], [0], job_id=1)

@@ -1736,8 +1710,7 @@ class TestFinishRequestServerSide:

1710

1711 def test_no_inflight_unmatched_demand_sends_failure(self):

1712 """finish_request with unmatched demand and no inflight ->

1739 - immediate TransferDoneMsg(success=False); _outbound cleared.

1740 - """

1713 + immediate TransferDoneMsg(success=False); _outbound cleared."""

1714 session, conn, _ = _make_session()

1715 _activate(session, conn)

1716 # Decoder demanded a block we never stored.

@@ -1763,8 +1736,7 @@ class TestFinishRequestServerSide:

1736

1737 def test_with_inflight_defers_then_fires_on_last_transfer(self):

1738 """finish_request with inflight defers; last transfer fires the

1766 - early-fail message and clears _outbound.

1767 - """

1739 + early-fail message and clears _outbound."""

1740 session, conn, transport = _make_session()

1741 _activate(session, conn)

1742 # Demand 2 blocks; we store 1 (kicks one inflight transfer).

@@ -1801,8 +1773,7 @@ class TestFinishRequestServerSide:

1773

1774 def test_full_demand_satisfied_still_sends_success(self):

1775 """finish_request must not override a fully-satisfied transfer:

1804 - when remaining hits 0, success=True still fires.

1805 - """

1776 + when remaining hits 0, success=True still fires."""

1777 session, conn, transport = _make_session()

1778 _activate(session, conn)

1779 conn.enqueue(

@@ -1831,8 +1802,7 @@ class TestFinishRequestServerSide:

1802 """Prefiller-first: finish_request runs before the decoder's

1803 fetch arrives. State is held until fetch, then

1804 finalized — success=True if all demand was matched against

1834 - available blocks, else success=False.

1835 - """

1805 + available blocks, else success=False."""

1806 # Case A: all demand satisfied by available blocks.

1807 session, conn, transport = _make_session()

1808 _activate(session, conn)

@@ -1893,8 +1863,7 @@ class TestFinishRequestServerSide:

1863 """finish_request with a stored-but-unmatched job and no inflight ->

1864 TransferDoneMsg(success=False) AND deferred

1865 StoreResult(success=False) for the submit_store'd job, instead of

1896 - the 30s _STORE_TIMEOUT_S path.

1897 - """

1866 + the 30s _STORE_TIMEOUT_S path."""

1867 session, conn, _ = _make_session()

1868 _activate(session, conn)

1869 # Decoder demanded b"demand"; we never stored it.

@@ -1930,8 +1899,7 @@ class TestFinishRequestServerSide:

1899 def test_finish_request_remaining_zero_emits_success_via_inflight(self):

1900 """Deferred-via-inflight path: finish_request with inflight, last

1901 transfer drains remaining to 0 -> TransferDoneMsg(success=True)

1933 - AND StoreResult(success=True).

1934 - """

1902 + AND StoreResult(success=True)."""

1903 session, conn, transport = _make_session()

1904 _activate(session, conn)

1905 conn.enqueue(

@@ -2279,8 +2247,7 @@ class TestDisconnect:

2247 def test_send_failure_marks_connection_dead(self):

2248 """A raising send must mark the connection dead, not silently drop

2249 the message — otherwise the session lingers alive, is never reaped,

2282 - and in-flight lookups/loads toward the dead peer hang forever.

2283 - """

2250 + and in-flight lookups/loads toward the dead peer hang forever."""

2251 session, conn, _ = _make_session()

2252 _activate(session, conn)

2253 assert session.alive

@@ -2293,8 +2260,7 @@ class TestDisconnect:

2260

2261 def test_close_surfaces_inflight_lookups(self):

2262 """close() reports kv_request_ids whose symmetric-P2P probe is still

2296 - unresolved; resolved probes are not reported (their answer is in).

2297 - """

2263 + unresolved; resolved probes are not reported (their answer is in)."""

2264 session, conn, _ = _make_session()

2265 _activate(session, conn)

2266

@@ -2402,8 +2368,7 @@ class TestDispatchErrorHandling:

2368

2369 def test_value_error_disconnects_on_first_occurrence(self):

2370 """A FetchMsg that fails validate() raises ValueError and must

2405 - terminate the session immediately, with a DisconnectMsg sent.

2406 - """

2371 + terminate the session immediately, with a DisconnectMsg sent."""

2372 session, conn, _ = _make_session()

2373 _activate(session, conn)

2374 # length mismatch → FetchMsg.validate raises ValueError

@@ -2430,8 +2395,7 @@ class TestDispatchErrorHandling:

2395

2396 def test_internal_error_does_not_disconnect_once(self):

2397 """A non-ValueError raised by a handler is treated as an internal

2433 - bug: counter increments, session stays alive on a single hit.

2434 - """

2398 + bug: counter increments, session stays alive on a single hit."""

2399 session, conn, _ = _make_session()

2400 _activate(session, conn)

2401

@@ -2454,8 +2418,7 @@ class TestDispatchErrorHandling:

2418

2419 def test_internal_error_threshold_disconnects(self):

2420 """Once consecutive non-protocol errors hit the threshold, the

2457 - session tears down via _protocol_error.

2458 - """

2421 + session tears down via _protocol_error."""

2422 session, conn, _ = _make_session()

2423 _activate(session, conn)

2424

@@ -2479,8 +2442,7 @@ class TestDispatchErrorHandling:

2442

2443 def test_internal_error_counter_resets_on_success(self):

2444 """A successful dispatch between errors prevents the threshold

2482 - from being reached.

2483 - """

2445 + from being reached."""

2446 session, conn, _ = _make_session()

2447 _activate(session, conn)

2448

@@ -2523,8 +2485,7 @@ class TestInflightPerReqInvariant:

2485 def test_invariant_holds_through_lifecycle(self):

2486 """Run a full submit→complete sequence for two concurrent

2487 kv_request_ids and assert the counter matches `_inflight` at

2526 - every observable step, including the empty-after-finish case.

2527 - """

2488 + every observable step, including the empty-after-finish case."""

2489 session, conn, transport = _make_session()

2490 _activate(session, conn)

2491

@@ -2589,8 +2550,7 @@ class TestInflightPerReqInvariant:

2550 """Populate many inflight xfers across many ids; lookup must

2551 match the actual presence in `_inflight` for both hits and

2552 misses. The whole point of the counter is that this lookup is

2592 - constant-time, but we assert correctness, not timing.

2593 - """

2553 + constant-time, but we assert correctness, not timing."""

2554 session, _, _ = _make_session()

2555 for kv_id_idx in range(100):

2556 kv_id = f"req-{kv_id_idx}"

tests/v1/kv_offload/tiering/test_async_lookup.py

+2 -4

@@ -264,8 +264,7 @@ class TestAsyncLookupManager:

264 subsequent lookup then returns False (MISS) directly, WITHOUT enqueuing

265 a fresh batch_lookup — that is what makes the request unable to loop.

266 The entry is retained (as False) not dropped, so cleanup()'s reverse

267 - index stays consistent; an unknown key is a no-op.

268 - """

267 + index stays consistent; an unknown key is a no-op."""

268 mgr = InMemoryLookupManager(existing_keys={_key(1), _key(2)})

269 ctx = _ctx("reqA")

270 assert mgr.lookup(_key(1), ctx) is None

@@ -305,8 +304,7 @@ class TestAsyncLookupManager:

304 receive at most one result per key. Normal operation resolves a key a

305 single time; a second result for an already-decided key trips the assert

306 that guards the invariant (a silent overwrite could flip a corrected

308 - miss back to True and reopen the failed-load livelock).

309 - """

307 + miss back to True and reopen the failed-load livelock)."""

308 mgr = InMemoryLookupManager(existing_keys={_key(1)})

309 ctx = _ctx("reqA")

310

tests/v1/kv_offload/tiering/test_fs_tier.py

+8 -16

@@ -369,8 +369,7 @@ def test_shutdown_discards_pending_tasks(fs_tier):

369 @pytest.mark.parametrize("use_c_ext", [True, False])

370 def test_store_load_data_integrity(fs_tier, monkeypatch, use_c_ext, batch_size):

371 """Data written by store must be exactly recovered by load, for batches

372 - of any size -- including the empty batch.

373 - """

372 + of any size -- including the empty batch."""

373 import vllm.v1.kv_offload.tiering.fs.io as io_mod

374

375 if use_c_ext and not io_mod._HAS_FSIO_C:

@@ -467,8 +466,7 @@ def test_wait_idle_blocks_until_tasks_complete():

466

467 def test_batch_lookup_c_extension(tmp_path):

468 """Validates batch_lookup_C: empty, single, all-existing, all-missing,

470 - mixed ordering, and input type validation.

471 - """

469 + mixed ordering, and input type validation."""

470 try:

471 from vllm.fs_io_C import batch_lookup as batch_lookup_C

472 except ImportError:

@@ -532,8 +530,7 @@ def test_batch_lookup_dispatch(fs_tier, monkeypatch, use_c_ext):

530 @pytest.mark.parametrize("use_c_ext", [True, False])

531 def test_out_of_bounds_block_id_smoke(fs_tier, monkeypatch, use_c_ext):

532 """Smoke test: a block id beyond the primary tensor's block count must

535 - fail the job, for both the C extension and the Python fallback.

536 - """

533 + fail the job, for both the C extension and the Python fallback."""

534 import vllm.v1.kv_offload.tiering.fs.io as io_mod

535

536 if use_c_ext and not io_mod._HAS_FSIO_C:

@@ -615,8 +612,7 @@ def test_batched_partial_load_failure_keeps_loaded_blocks(

612 loaded before it (#50321). Corrupt the LAST block: the earlier blocks load

613 fine, so the job reports successful_keys for them and marks only the failed

614 tail a miss. The earlier keys stay HIT — including for the same request —

618 - while the corrupt block stays a MISS (its file was removed).

619 - """

615 + while the corrupt block stays a MISS (its file was removed)."""

616 import vllm.v1.kv_offload.tiering.fs.io as io_mod

617

618 if use_c_ext and not io_mod._HAS_FSIO_C:

@@ -668,8 +664,7 @@ def test_batched_load_first_block_fails_marks_whole_batch(

664 ):

665 """When the FIRST block fails, nothing loaded before it: the job reports no

666 successful_keys (None) and the whole batch is marked a miss for the

671 - request.

672 - """

667 + request."""

668 import vllm.v1.kv_offload.tiering.fs.io as io_mod

669

670 if use_c_ext and not io_mod._HAS_FSIO_C:

@@ -700,8 +695,7 @@ def test_transient_load_failure_leaves_file(fs_tier, monkeypatch, use_c_ext):

695 """A transient host error (here ELOOP on open) is NOT a short read: the job

696 fails but the block file must survive untouched, on both the C and Python

697 paths. Deleting on a transient error would turn a passing hiccup into

703 - permanent data loss.

704 - """

698 + permanent data loss."""

699 import vllm.v1.kv_offload.tiering.fs.io as io_mod

700

701 if use_c_ext and not io_mod._HAS_FSIO_C:

@@ -805,8 +799,7 @@ def test_mixed_job_results_emit_event_only_for_successful_job(

799 fs_tier_with_events, monkeypatch

800 ):

801 """With a failed and a successful store job in flight, exactly one event

808 - is emitted and its keys belong to the successful job.

809 - """

802 + is emitted and its keys belong to the successful job."""

803 import vllm.v1.kv_offload.tiering.fs.manager as mgr_mod

804

805 tier = fs_tier_with_events

@@ -891,8 +884,7 @@ def test_events_require_global_kv_events_flag(tmp_path):

884

885 def test_cascade_store_emits_fs_event_through_tiering_manager(tmp_path):

886 """A GPU->CPU->fs cascade surfaces the tier-owned FS stored event via the

894 - TieringOffloadingManager's aggregated take_events().

895 - """

887 + TieringOffloadingManager's aggregated take_events()."""

888 from vllm.v1.kv_offload.tiering.manager import (

889 CPUPrimaryTierOffloadingManager,

890 TieringOffloadingManager,

tests/v1/kv_offload/tiering/test_obj_tier.py

+2 -4

@@ -346,8 +346,7 @@ class TestMockObjTierBasic:

346 the scheduler would re-issue the same doomed promotion every step for

347 the life of the request. The mark is served from cache with no

348 re-probe, so even though the mock object is still 'present' the SAME

349 - request now resolves to MISS.

350 - """

349 + request now resolves to MISS."""

350 ctx = ReqContext(req_id="obj-livelock")

351 self.tier.submit_store(make_job(1, [key(1)], [0]))

352 assert all(r.success for r in drain(self.tier))

@@ -651,8 +650,7 @@ class TestObjTierKVEvents:

650 def test_mixed_job_results_emit_event_only_for_successful_job(self):

651 """With a failed and a successful store job resolving in the same

652 poll, exactly one event is emitted and its keys belong to the

654 - successful job.

655 - """

653 + successful job."""

654 original = self.agent.check_xfer_state

655 self.agent.check_xfer_state = lambda h: "ERR" if h._id == 0 else original(h)

656 self.tier.submit_store(make_job(1, [key(1)], [0])) # handle 0: fails

tests/v1/kv_offload/tiering/test_tiering_offloading.py

+11 -22

@@ -273,8 +273,7 @@ _CASCADE_SUPPLY_BEHAVIOR = {

273

274 def test_every_lookup_result_has_a_declared_cascade_behavior():

275 """A new LookupResult must state what the cascade does with it, rather

276 - than falling into whichever branch happens to catch it.

277 - """

276 + than falling into whichever branch happens to catch it."""

277 assert set(_CASCADE_SUPPLY_BEHAVIOR) == set(LookupResult)

278

279

@@ -321,8 +320,7 @@ class TestTieringOffloadingManager:

320

321 def test_failed_promotion_finalizes_primary_with_failure(self, manager_setup):

322 """A failed promotion still finalizes the primary slots with

324 - success=False; the tier corrects its own verdict.

325 - """

323 + success=False; the tier corrects its own verdict."""

324 from unittest.mock import patch

325

326 from vllm.v1.kv_offload.tiering.base import JobResult

@@ -1093,8 +1091,7 @@ class TestTieringOffloadingManager:

1091

1092 def test_on_new_request_lifecycle(self, manager_setup):

1093 """Policy defaults to CHUNK_LEVEL, escalates when a tier requests it,

1096 - and is cleaned up on on_request_finished.

1097 - """

1094 + and is cleaned up on on_request_finished."""

1095 # Default: all tiers return CHUNK_LEVEL

1096 ctx = ReqContext(req_id="req_policy_lifecycle")

1097 result = self.manager.on_new_request(ctx)

@@ -1171,8 +1168,7 @@ class TestTieringOffloadingManager:

1168

1169 def _make_request_level_request(self, req_id: str) -> ReqContext:

1170 """Start a request for which tier1 asks for request-level offloading,

1174 - with tier1's submit_store wrapped so the cascade can be observed.

1175 - """

1171 + with tier1's submit_store wrapped so the cascade can be observed."""

1172 self.secondary_tier1.on_new_request = lambda req_context: (

1173 RequestOffloadingContext(policy=OffloadPolicy.REQUEST_LEVEL)

1174 )

@@ -1194,8 +1190,7 @@ class TestTieringOffloadingManager:

1190 def test_cascade_rejects_retry_from_primary(self, manager_setup):

1191 """RETRY would be parked with no guarantee of ever draining, so the

1192 cascade refuses it. No real primary tier returns it, so this is the one

1197 - disposition that has to be forced.

1198 - """

1193 + disposition that has to be forced."""

1194 keys = to_keys(range(3))

1195 self._start_request()

1196 assert self.manager.prepare_store(keys, _CTX) is not None

@@ -1210,8 +1205,7 @@ class TestTieringOffloadingManager:

1205

1206 def _start_in_flight_primary_write(self, keys: list[OffloadKey]) -> ReqContext:

1207 """Leave a GPU->primary write for `keys` open, so they look present to

1213 - prepare_store but are not yet readable.

1214 - """

1208 + prepare_store but are not yet readable."""

1209 writer_ctx = ReqContext(req_id="req_writer")

1210 self._start_request(writer_ctx)

1211 assert self.manager.prepare_store(keys, writer_ctx) is not None

@@ -1220,8 +1214,7 @@ class TestTieringOffloadingManager:

1214 def test_cascade_defers_keys_whose_primary_write_is_in_flight(self, manager_setup):

1215 """A key another request is still writing must reach the peer once the

1216 write lands, not be dropped: prepare_store already counts it as stored

1223 - and the scheduler advances past its chunk, so nothing re-offers it.

1224 - """

1217 + and the scheduler advances past its chunk, so nothing re-offers it."""

1218 keys = to_keys(range(3))

1219 writer_ctx = self._start_in_flight_primary_write(keys)

1220

@@ -1238,8 +1231,7 @@ class TestTieringOffloadingManager:

1231

1232 def test_deferred_cascade_holds_request_from_finalization(self, manager_setup):

1233 """A request that finishes with keys still deferred must stay alive, or

1241 - its tiers are torn down before the peer is ever served.

1242 - """

1234 + its tiers are torn down before the peer is ever served."""

1235 keys = to_keys(range(3))

1236 writer_ctx = self._start_in_flight_primary_write(keys)

1237

@@ -1262,8 +1254,7 @@ class TestTieringOffloadingManager:

1254 self, manager_setup

1255 ):

1256 """A failed write frees the chunk, so the deferred key resolves to MISS

1265 - and the request finalizes instead of parking forever.

1266 - """

1257 + and the request finalizes instead of parking forever."""

1258 keys = to_keys(range(3))

1259 writer_ctx = self._start_in_flight_primary_write(keys)

1260

@@ -1283,8 +1274,7 @@ class TestTieringOffloadingManager:

1274 def test_reset_cache_clears_orchestrator_state(self, manager_setup):

1275 """reset_cache wipes every kind of orchestrator state and resets

1276 primary tier; pending submissions are dropped without being sent

1286 - to the secondary tier. Active request state is retained.

1287 - """

1277 + to the secondary tier. Active request state is retained."""

1278 # Cascade — populates primary chunks and leaves cascade jobs

1279 # in _jobs (the synchronous example tier has already

1280 # queued completions); reset_cache's drain loop will pick them up.

@@ -1376,8 +1366,7 @@ class TestTieringOffloadingManager:

1366 self, manager_setup, load_tier_filter

1367 ):

1368 """Filter excluding secondary medium returns MISS from secondaries

1379 - even when they hold the chunk; primary is unaffected.

1380 - """

1369 + even when they hold the chunk; primary is unaffected."""

1370 chunks = to_keys(range(2))

1371 # Put one chunk in primary, one only in secondary

1372 self._start_request()

tests/v1/logits_processors/test_correctness.py

+1 -2

@@ -1344,8 +1344,7 @@ def test_thinking_budget_invalid_budget_rejected(invalid_budget):

1344

1345 def test_thinking_budget_long_thinking_section_end_marker_found_at_correct_index():

1346 """Test thinking budget enforced for a long thinking run,

1347 - then a natural end marker.

1348 - """

1347 + then a natural end marker."""

1348 h = ThinkingBudgetStateHolder(

1349 MockReasoningConfig(), 8, 0, torch.device("cpu"), False

1350 )

tests/v1/logits_processors/utils.py

+2 -4

@@ -138,8 +138,7 @@ class EntryPoints(list):

138

139 class DummyPerReqLogitsProcessor:

140 """The request-level logits processor masks out all logits except the

141 - token id identified by `target_token`

142 - """

141 + token id identified by `target_token`"""

142

143 def __init__(self, target_token: int) -> None:

144 """Specify `target_token`."""

@@ -158,8 +157,7 @@ class DummyPerReqLogitsProcessor:

157

158 class WrappedPerReqLogitsProcessor(AdapterLogitsProcessor):

159 """Example of wrapping a fake request-level logit processor to create a

161 - batch-level logits processor

162 - """

160 + batch-level logits processor"""

161

162 def is_argmax_invariant(self) -> bool:

163 return False

tests/v1/metrics/test_perf_metrics.py

+1 -2

@@ -1241,8 +1241,7 @@ def test_mla_per_gpu_with_pipeline_parallelism():

1241

1242 def test_mla_model_metrics_excludes_standard_attention():

1243 """Test that ModelMetrics uses MLAAttentionMetrics, not AttentionMetrics,

1244 - for DeepSeek MLA models.

1245 - """

1244 + for DeepSeek MLA models."""

1245 hf_config = DeepseekV3Config(

1246 hidden_size=7168,

1247 num_attention_heads=128,

tests/v1/metrics/test_ray_metrics.py

+7 -14

@@ -45,8 +45,7 @@ def test_engine_log_metrics_ray(

45 max_tokens: int,

46 ) -> None:

47 """Simple smoke test, verifying this can be used without exceptions.

48 - Need to start a Ray cluster in order to verify outputs.

49 - """

48 + Need to start a Ray cluster in order to verify outputs."""

49

50 @ray.remote(num_gpus=1)

51 class EngineTestActor:

@@ -120,8 +119,7 @@ def test_sanitized_opentelemetry_name():

119 def _install_mock_metric(wrapper: RayPrometheusMetric) -> MagicMock:

120 """Swap the wrapper's underlying Ray metric for a MagicMock while

121 preserving the real metric's ``_tag_keys`` (labels() reads them to

123 - validate arity).

124 - """

122 + validate arity)."""

123 real_metric = wrapper.metric

124 mock = MagicMock()

125 mock._tag_keys = real_metric._tag_keys

@@ -131,8 +129,7 @@ def _install_mock_metric(wrapper: RayPrometheusMetric) -> MagicMock:

129

130 def test_ray_counter_labels_returns_independent_children():

131 """RayCounterWrapper.labels() must return distinct labeled children that

134 - each carry their own tag set.

135 - """

132 + each carry their own tag set."""

133 base = RayCounterWrapper(

134 name="vllm_test_finish_reason",

135 documentation="",

@@ -152,8 +149,7 @@ def test_ray_counter_labels_returns_independent_children():

149

150 def test_ray_counter_inc_forwards_per_child_tags():

151 """.inc() on a labeled counter must forward that child's tags to the

155 - underlying Ray metric (not rely on a shared set_default_tags).

156 - """

152 + underlying Ray metric (not rely on a shared set_default_tags)."""

153 wrapper = RayCounterWrapper(

154 name="vllm_test_counter_tag_forward",

155 documentation="",

@@ -218,8 +214,7 @@ def test_ray_histogram_labels_returns_independent_children_and_forwards_tags():

214 def test_ray_counter_labels_accepts_non_string_label_values():

215 """RayPrometheusStatLogger passes ``str(idx)`` for engine indexes; this

216 covers the coercion path for any caller that passes a non-string label

221 - value positionally.

222 - """

217 + value positionally."""

218 wrapper = RayCounterWrapper(

219 name="vllm_test_nonstr_label",

220 documentation="",

@@ -242,8 +237,7 @@ def test_ray_counter_labels_arity_validation():

237

238 def test_unlabeled_inc_carries_replica_id():

239 """Recording on an unlabeled metric must still pass ReplicaId — it's a

245 - declared tag_key and Ray rejects updates that omit any declared key.

246 - """

240 + declared tag_key and Ray rejects updates that omit any declared key."""

241 wrapper = RayCounterWrapper(

242 name="vllm_test_unlabeled_replica_id",

243 documentation="",

@@ -256,8 +250,7 @@ def test_unlabeled_inc_carries_replica_id():

250

251 def test_double_labels_raises():

252 """labels() on an already-labeled child should raise, mirroring the

259 - prometheus_client contract.

260 - """

253 + prometheus_client contract."""

254 wrapper = RayCounterWrapper(

255 name="vllm_test_double_labels",

256 documentation="",

tests/v1/sample/test_batched_count_greater_than.py

+1 -2

@@ -1,8 +1,7 @@

1 # SPDX-License-Identifier: Apache-2.0

2 # SPDX-FileCopyrightText: Copyright contributors to the vLLM project

3 """Test that batched_count_greater_than does not trigger 0/1 specialization

4 -recompiles when batch_size varies.

5 -"""

4 +recompiles when batch_size varies."""

5

6 import torch

7

tests/v1/sample/test_logprobs.py

+1 -2

@@ -573,8 +573,7 @@ def test_logprobs_mode(logprobs_mode: LogprobsMode):

573 def test_prompt_logprobs_mode():

574 """prompt_logprobs must respect logprobs_mode: *_logits and *_logprobs

575 must return different values. Prompt tokens skip sampling processors,

576 - so processed_* == raw_* on the prompt side.

577 - """

576 + so processed_* == raw_* on the prompt side."""

577 from vllm import LLM

578

579 values: dict[str, float] = {}

tests/v1/sample/test_rejection_sampler.py

+1 -2

@@ -54,8 +54,7 @@ def create_logits_tensor(

54 token_idx_to_override: int | None = None,

55 ) -> torch.Tensor:

56 """Helper function to create logits tensor that

57 - will produce desired token ids on argmax

58 - """

57 + will produce desired token ids on argmax"""

58 token_ids = [tokens[:-1] for tokens in output_token_ids]

59 num_total_tokens = sum(len(tokens) for tokens in token_ids)

60 logits = torch.full((num_total_tokens, vocab_size), -100.0, device=DEVICE_TYPE)

tests/v1/sample/test_topk_topp_sampler.py

+1 -2

@@ -884,8 +884,7 @@ class TestFlashInferTopkToppRobustness:

884

885 def _make_logits(self, pattern: str) -> torch.Tensor:

886 """Build (BATCH, VOCAB) logits with `pattern` applied to row 0

887 - (rows 1..B-1 stay clean so we can detect cross-row corruption).

888 - """

887 + (rows 1..B-1 stay clean so we can detect cross-row corruption)."""

888 logits = (

889 torch.randn(

890 self.BATCH,

tests/v1/sample/utils.py

+1 -2

@@ -204,8 +204,7 @@ def fake_update_logitsprocs_state(

204 batch_update: BatchUpdate | None,

205 ) -> None:

206 """Imitate logits processors persistent batch state update

207 - in engine core

208 - """

207 + in engine core"""

208 for logitproc in test_fakes.get_logitsprocs():

209 logitproc.update_state(batch_update)

210 holder = test_fakes.sampling_metadata.thinking_budget_state_holder

tests/v1/simple_kv_offload/test_kv_events.py

+3 -6

@@ -67,8 +67,7 @@ def _make_mixed_kv_cache_config(

67 mamba_block_size: int = BLOCK_SIZE,

68 ) -> KVCacheConfig:

69 """Build a KVCacheConfig with a FullAttention group plus optional

70 - SlidingWindow and Mamba groups for per-group metadata tests.

71 - """

70 + SlidingWindow and Mamba groups for per-group metadata tests."""

71 register_all_kvcache_specs(vllm_config=None)

72 groups = []

73 tensors = []

@@ -239,8 +238,7 @@ def test_events_disabled_no_events() -> None:

238 # ---------------------------------------------------------------------------

239 def test_block_stored_per_group_metadata_full_attention() -> None:

240 """BlockStored carries medium, group_idx, kind, sliding_window, block_size,

242 - and the correct block hash.

243 - """

241 + and the correct block hash."""

242 fix = make_events_scheduler()

243 sched = fix.scheduler

244 assert sched.kv_event_medium == MEDIUM_CPU

@@ -640,8 +638,7 @@ def test_mamba_block_size_not_scaled() -> None:

638 def test_multi_group_fa_block_stored_group_idx() -> None:

639 """Two distinct groups (FA + SW): each BlockStored carries its group_idx

640 and kind. HybridKVCacheCoordinator requires >=2 distinct spec groups, so

643 - we use FA (group 0) + SlidingWindow (group 1).

644 - """

641 + we use FA (group 0) + SlidingWindow (group 1)."""

642 kv_cfg = _make_mixed_kv_cache_config(num_blocks=16, sliding_window=32)

643 fix = make_events_scheduler(

644 num_cpu_blocks=8,

tests/v1/simple_kv_offload/test_scheduler.py

+6 -12

@@ -1170,8 +1170,7 @@ def test_touched_blocks_survive_eviction() -> None:

1170 # ---------------------------------------------------------------------------

1171 def test_preemption_no_cpu_block_leak() -> None:

1172 """request_finished during in-flight load defers cleanup;

1173 - completes after load done.

1174 - """

1173 + completes after load done."""

1174 fix = make_scheduler(num_cpu_blocks=8, num_gpu_blocks=16, lazy=False)

1175 sched = fix.scheduler

1176

@@ -1266,8 +1265,7 @@ def test_eager_store_preemption_cleanup() -> None:

1265 # ---------------------------------------------------------------------------

1266 def test_inflight_finish_deferred_cleanup() -> None:

1267 """Store, then start a load, request_finished defers,

1269 - load completion fires cleanup.

1270 - """

1268 + load completion fires cleanup."""

1269 fix = make_scheduler(num_cpu_blocks=8, num_gpu_blocks=16, lazy=False)

1270 sched = fix.scheduler

1271

@@ -1401,8 +1399,7 @@ def test_chunked_prefill_reads_live_block_ids() -> None:

1399 """With chunked prefill, block IDs accumulate across scheduler steps.

1400 _prepare_eager_store_specs reads block IDs from scheduler_output via

1401 yield_req_data, so the store should reflect the updated (larger) block

1404 - list, not a stale snapshot.

1405 - """

1402 + list, not a stale snapshot."""

1403 fix = make_scheduler(num_cpu_blocks=8, num_gpu_blocks=16, lazy=False)

1404 sched = fix.scheduler

1405

@@ -1973,8 +1970,7 @@ def test_cp_eager_store_and_load_roundtrip(

1970 ) -> None:

1971 """With CP enabled, store blocks to CPU and reload them for a new request

1972 with matching tokens. Verifies that hash matching and transfer-pair

1976 - construction work with the virtual block size.

1977 - """

1973 + construction work with the virtual block size."""

1974 fix = _make_cp_scheduler(

1975 dcp_world_size=dcp_world_size,

1976 pcp_world_size=pcp_world_size,

@@ -2118,8 +2114,7 @@ def test_cp_effective_block_size_store_and_load(

2114 @pytest.mark.parametrize("cp_world_size", [1, 2, 4])

2115 def test_cp_lazy_target_blocks_scaling(cp_world_size: int) -> None:

2116 """_estimate_lazy_target_blocks returns fewer blocks when cp_world_size > 1

2121 - because each virtual block covers more tokens.

2122 - """

2117 + because each virtual block covers more tokens."""

2118 kv_cache_config = _make_kv_cache_config(num_blocks=16)

2119 max_batched = 64

2120

@@ -2811,8 +2806,7 @@ def test_finished_eager_store_recovers_nulled_window_tail() -> None:

2806 """A sliding-window group nulls pages that left the window before the

2807 connector sees the finished request's block table, while the GPU keeps

2808 them hashed in its free queue. The eager store must copy those pages, or

2814 - the CPU lookup (which needs the whole window) misses every time.

2815 - """

2809 + the CPU lookup (which needs the whole window) misses every time."""

2810 fix = make_scheduler(num_cpu_blocks=16, num_gpu_blocks=32, num_groups=2, lazy=False)

2811 sched = fix.scheduler

2812 gpu_pool = fix.gpu_block_pool

tests/v1/spec_decode/test_backup_token_async_spec.py

+1 -2

@@ -131,8 +131,7 @@ class TestBackupTokenAsyncSpec:

131

132 def test_buggy_code_was_always_off_by_one(self):

133 """The original code used seq_len as index, which is always one past

134 - the end of output_token_ids even without async inflation.

135 - """

134 + the end of output_token_ids even without async inflation."""

135 req_ids = ["r0"]

136 requests = {"r0": _FakeRequest([0, 1, 2], [1000, 1001])}

137 batch = _FakeInputBatch(req_ids, [5])

tests/v1/spec_decode/test_eagle_draft_attn_metadata.py

+3 -6

@@ -105,8 +105,7 @@ def test_build_draft_attn_metadata_sets_seq_lens_cpu_upper_bound():

105 """The fix: every per-step ``CommonAttentionMetadata`` carries a non-None

106 ``seq_lens_cpu_upper_bound`` derived from the target-side upper bound plus

107 the current draft-step offset. Padded entries are zeroed (matching the

108 - main model runner's convention).

109 - """

108 + main model runner's convention)."""

109 fake = _make_fake_speculator()

110 base = torch.tensor([100, 200, 300, 0], dtype=torch.int32)

111

@@ -127,8 +126,7 @@ def test_build_draft_attn_metadata_sets_seq_lens_cpu_upper_bound():

126

127 def test_build_draft_attn_metadata_zeroes_padded_upper_bound_tail():

128 """The padded tail of the upper-bound tensor is zeroed, so it stays a

130 - valid tensor of length ``num_reqs_padded`` regardless of padding.

131 - """

129 + valid tensor of length ``num_reqs_padded`` regardless of padding."""

130 fake = _make_fake_speculator()

131 base = torch.zeros(2, dtype=torch.int32)

132

@@ -143,8 +141,7 @@ def test_build_draft_attn_metadata_zeroes_padded_upper_bound_tail():

141

142 def test_build_draft_attn_metadata_clamps_to_max_model_len():

143 """The per-request upper bound (target bound + step) is clamped to the

146 - model length so it never exceeds the allocated KV range.

147 - """

144 + model length so it never exceeds the allocated KV range."""

145 fake = _make_fake_speculator(max_model_len=1024)

146 base = torch.tensor([1023, 500], dtype=torch.int32)

147

tests/v1/spec_decode/test_eagle_step_kernel.py

+1 -2

@@ -132,8 +132,7 @@ def test_eagle_step_slot_mapping_kernel_exceeds_max():

132

133 def test_eagle_step_slot_mapping_kernel_cudagraph_padding():

134 """Test that padding threads write PADDING_SLOT_ID when

135 - input_batch_size > batch_size (cudagraph padding).

136 - """

135 + input_batch_size > batch_size (cudagraph padding)."""

136 device = torch.device(DEVICE_TYPE)

137 batch_size = 4

138 input_batch_size = 8

tests/v1/spec_decode/test_llm_base_proposer.py

+2 -4

@@ -69,8 +69,7 @@ def _make_kv_cache_config(layer_names: set[str]) -> SimpleNamespace:

69

70 def test_block_size_uses_kernel_block_size(monkeypatch: pytest.MonkeyPatch):

71 """The proposer's slot-mapping math runs against the kernel-granularity

72 - block table, so block_size must come from kernel_block_sizes.

73 - """

72 + block table, so block_size must come from kernel_block_sizes."""

73 layer_names = {"draft.0.self_attn.attn"}

74 proposer = _make_proposer(monkeypatch, layer_names)

75

@@ -98,8 +97,7 @@ def test_block_size_falls_back_to_kv_cache_spec(monkeypatch: pytest.MonkeyPatch)

97

98 def test_draft_layer_iteration_is_deterministic(monkeypatch: pytest.MonkeyPatch):

99 """_draft_attn_layer_names is a set; the attention groups built from it

101 - must not depend on its (process-random) iteration order.

102 - """

100 + must not depend on its (process-random) iteration order."""

101 layer_names = {"draft.c.attn", "draft.a.attn", "draft.b.attn"}

102 expected_order = sorted(layer_names)

103

tests/v1/spec_decode/test_max_len.py

+1 -2

@@ -93,8 +93,7 @@ def test_eagle_max_len(

93 @pytest.mark.parametrize("spec_max_model_len", [80, 150])

94 def test_mtp_speculative_config_max_model_len(spec_max_model_len: int):

95 """Regression test for #41456: max_model_len in speculative config

96 - should be respected for the draft model.

97 - """

96 + should be respected for the draft model."""

97 model_config = ModelConfig(

98 model="XiaomiMiMo/MiMo-7B-Base",

99 runner="generate",

tests/v1/structured_output/test_validation.py

+4 -8

@@ -29,8 +29,7 @@ class _StubModelConfig:

29 def test_structured_outputs_rejected_for_diffusion_models():

30 """Diffusion LLMs denoise the canvas in parallel, which is incompatible

31 with the token-by-token grammar FSM. The request must fail with a clear

32 - validation error instead of an FSM rejection mid-generation (#45436).

33 - """

32 + validation error instead of an FSM rejection mid-generation (#45436)."""

33 params = SamplingParams(

34 structured_outputs=StructuredOutputsParams(json=JSON_SCHEMA)

35 )

@@ -63,8 +62,7 @@ def test_degenerate_structured_outputs_rejected(structured_outputs, match):

62 """json_object=False and an empty json schema pass the `is not None`

63 exclusivity check but resolve to no structured-output key, so they must be

64 rejected at request validation (-> 400) instead of reaching and crashing

66 - the engine.

67 - """

65 + the engine."""

66 params = SamplingParams(structured_outputs=structured_outputs)

67 with pytest.raises(VLLMValidationError, match=match):

68 params._validate_structured_outputs(

@@ -87,8 +85,7 @@ def test_regex_with_nul_byte_rejected(regex):

85 handled by xgrammar's native regex converter. It must be rejected at request

86 validation in every backend mode (a clean 400), instead of reaching that

87 native code or silently falling back to another backend in the default

90 - 'auto' mode.

91 - """

88 + 'auto' mode."""

89 params = SamplingParams(structured_outputs=StructuredOutputsParams(regex=regex))

90

91 # Rejected before backend selection, so it is a 400 even in 'auto' mode

@@ -127,8 +124,7 @@ INVALID_JSON_SCHEMA = {"type": "object", "properties": {"name": {"type": "str"}}

124 )

125 def test_unsupported_grammar_is_a_client_error(backend, structured_outputs):

126 """Only `VLLMClientError` survives `AsyncLLM.generate` untouched; anything else

130 - is wrapped in `EngineGenerateError` and served as a 500 instead of a 400.

131 - """

127 + is wrapped in `EngineGenerateError` and served as a 500 instead of a 400."""

128 params = SamplingParams(structured_outputs=structured_outputs)

129 with pytest.raises(VLLMClientError):

130 params._validate_structured_outputs(

tests/v1/test_outputs.py

+4 -8

@@ -46,8 +46,7 @@ def test_logprobs_tensors_cat():

46 def test_logprobs_tensors_tolists_with_tensor_boundaries():

47 """Adaptive verification hands over the request boundaries as a tensor

48 (they only exist on device); tolists() must materialize it as a plain

49 - list so slice_request splits requests correctly.

50 - """

49 + list so slice_request splits requests correctly."""

50 tensors = LogprobsTensors(

51 torch.tensor([[1, 2], [3, 4], [5, 6]]),

52 torch.tensor([[0.1, 0.2], [0.3, 0.4], [0.5, 0.6]]),

@@ -86,8 +85,7 @@ def test_sampling_mask_lists_to_nested_list():

85 def test_sampling_mask_tensors_match_finite_support(max_num_kept):

86 """Whatever its size (empty, one, around the compact width, beyond the

87 cap, the whole vocab), a sampled row's mask is exactly the ascending set

89 - of its finite logits; rows that sampled nothing are empty.

90 - """

88 + of its finite logits; rows that sampled nothing are empty."""

89 from vllm.v1.worker.gpu.sample.output import MAX_COMPACT_SUPPORT

90

91 vocab_size = 20_001

@@ -114,8 +112,7 @@ def test_sampling_mask_tensors_match_finite_support(max_num_kept):

112 def test_sampling_mask_matches_processed_top_k_top_p_support():

113 """The mask must exactly mirror whatever support `apply_top_k_top_p`

114 (the real logits-processing function used by the sampler) actually

117 - produces, whatever backend implements it.

118 - """

115 + produces, whatever backend implements it."""

116 processed_logits = apply_top_k_top_p(

117 logits=torch.tensor(

118 [[6.0, 5.0, 4.0, 4.0, 4.0, 2.0, 1.0, 0.0]], device=DEVICE_TYPE

@@ -141,8 +138,7 @@ def test_sampling_mask_matches_processed_top_k_top_p_support():

138 def test_sampling_mask_preserves_top_k_boundary_ties():

139 """When the kept support is wider than `max_num_kept` (e.g. a top-k

140 boundary tie keeps more than k logits), the mask must fall back to the

144 - exact bitmask instead of silently truncating to `max_num_kept` ids.

145 - """

141 + exact bitmask instead of silently truncating to `max_num_kept` ids."""

142 processed_logits = torch.tensor(

143 [[6.0, 5.0, 4.0, 4.0, 4.0, float("-inf"), float("-inf"), float("-inf")]],

144 device=DEVICE_TYPE,

tests/v1/test_serial_utils.py

+1 -2

@@ -427,8 +427,7 @@ def _logprobs_outputs(num_reqs: int, num_prompt_tokens: int):

427 """An EngineCoreOutputs carrying prompt logprobs, as the engine core sends

428 it: many requests, each with per-token tensors small enough that pyzmq

429 copies their frames, while the accumulated payload frame is large enough

430 - that pyzmq sends it zero-copy.

431 - """

430 + that pyzmq sends it zero-copy."""

431 from vllm.v1.engine import EngineCoreOutput, EngineCoreOutputs

432 from vllm.v1.outputs import LogprobsTensors

433

tests/v1/worker/test_dsv4_packed_zeroer_geometry.py

+1 -2

@@ -101,8 +101,7 @@ def test_packed_dsv4_zeroer_zeroes_only_each_layers_page():

101 def test_overlaid_zeroer_dedups_segments_with_max_span():

102 """Two groups overlay one allocation; the zeroer must emit one segment per distinct

103 byte offset, spanning the widest overlaid page, so a newly allocated block is fully

104 - zeroed no matter which group owns it.

105 - """

104 + zeroed no matter which group owns it."""

105 from unittest.mock import MagicMock

106

107 from vllm.v1.core.kv_cache_utils import get_kv_cache_config_from_groups

tests/v1/worker/test_encoder_runner.py

+5 -10

@@ -97,8 +97,7 @@ def test_draft_lookahead_uses_boundary_feature_when_cached():

97 """The drafter's +1 look-ahead can reach the feature at offset ==

98 processed_end (the next chunk). When its encoder output is already cached

99 (the scheduler encoded it ahead), it is used for the look-ahead position

100 - rather than ignored.

101 - """

100 + rather than ignored."""

101 f0 = _feature("h0", offset=0, length=8)

102 f1 = _feature("h1", offset=8, length=8) # starts exactly at processed_end

103 runner = _make_runner([f0, f1], cached=[f0, f1])

@@ -115,8 +114,7 @@ def test_draft_lookahead_uses_boundary_feature_when_cached():

114 def test_draft_lookahead_tolerates_missing_boundary_feature():

115 """When the +1 look-ahead feature past the processed boundary is not yet

116 encoded, fall back to the token embedding (the draft token is verified by

118 - the target) instead of raising.

119 - """

117 + the target) instead of raising."""

118 f0 = _feature("h0", offset=0, length=8)

119 f1 = _feature("h1", offset=8, length=8) # boundary feature, not cached

120 runner = _make_runner([f0, f1], cached=[f0])

@@ -133,8 +131,7 @@ def test_draft_lookahead_tolerates_missing_boundary_feature():

131 def test_draft_lookahead_raises_on_interior_miss():

132 """A miss for a feature within the processed range (not the look-ahead

133 boundary) is a real invariant violation and must fail loudly, even on the

136 - drafter path.

137 - """

134 + drafter path."""

135 f0 = _feature("h0", offset=0, length=8) # interior, within processed range

136 runner = _make_runner([f0], cached=[])

137

@@ -144,8 +141,7 @@ def test_draft_lookahead_raises_on_interior_miss():

141

142 def test_target_path_raises_on_encoder_cache_miss():

143 """On the target path (no look-ahead) a miss is a real invariant

147 - violation and must fail loudly.

148 - """

144 + violation and must fail loudly."""

145 f0 = _feature("h0", offset=0, length=8)

146 runner = _make_runner([f0], cached=[])

147

@@ -156,8 +152,7 @@ def test_target_path_raises_on_encoder_cache_miss():

152 @pytest.mark.parametrize("draft_lookahead", [0, 1])

153 def test_multi_request_batch_gathers_per_request(draft_lookahead):

154 """Two prefilling requests in one batch: per-request query bounds must be

159 - indexed by request, not applied as whole arrays.

160 - """

155 + indexed by request, not applied as whole arrays."""

156 a0 = _feature("a0", offset=0, length=8)

157 b0 = _feature("b0", offset=0, length=8)

158 cache = EncoderCache()

tests/v1/worker/test_gpu_bad_words.py

+3 -6

@@ -73,8 +73,7 @@ def _apply(bad_words_token_ids: list[list[int]]) -> torch.Tensor:

73

74 def test_v2_bad_words_prefix_inside_draft_tokens():

75 """A prefix matching entirely within the draft tokens must mask the bad

76 - word's last token at the draft position that completes the prefix.

77 - """

76 + word's last token at the draft position that completes the prefix."""

77 out = _apply([[12, 13, 40]])

78 expected = torch.zeros_like(out)

79 expected[2, 40] = -float("inf")

@@ -83,8 +82,7 @@ def test_v2_bad_words_prefix_inside_draft_tokens():

82

83 def test_v2_bad_words_prefix_spanning_committed_and_draft_tokens():

84 """A prefix spanning the committed/draft boundary must mask at the row

86 - where the prefix completes, not one draft position later.

87 - """

85 + where the prefix completes, not one draft position later."""

86 out = _apply([[11, 12, 30]])

87 expected = torch.zeros_like(out)

88 expected[1, 30] = -float("inf")

@@ -93,8 +91,7 @@ def test_v2_bad_words_prefix_spanning_committed_and_draft_tokens():

91

92 def test_v2_bad_words_no_spurious_match_from_last_committed_token():

93 """The last committed token must not be double-counted as the first draft

96 - token; [11, 11] never occurs in output [10, 11] + drafts [12, 13].

97 - """

94 + token; [11, 11] never occurs in output [10, 11] + drafts [12, 13]."""

95 out = _apply([[11, 11, 50]])

96 expected = torch.zeros_like(out)

97 torch.testing.assert_close(out, expected)

tests/v1/worker/test_gpu_batch_shard.py

+9 -18

@@ -43,8 +43,7 @@ def _make_batch(

43 slot_pool: np.ndarray | None = None,

44 ) -> InputBatch:

45 """Consistent batch with random unique request slots, 1 + k_i logits

46 - rows per request at the tail of each query segment.

47 - """

46 + rows per request at the tail of each query segment."""

47 if slot_pool is None:

48 slot_pool = np.arange(max_num_reqs)

49 slots = rng.choice(slot_pool, size=num_reqs, replace=False)

@@ -142,8 +141,7 @@ def _owned_rows(cu: np.ndarray, owned: np.ndarray) -> np.ndarray:

141 def test_local_batch_partition(tp_size: int, seed: int):

142 """Local sub-batches partition the batch by owner, preserve batch order

143 within an owner, and carry exactly the owned requests' entries of every

145 - per-request and per-row array.

146 - """

144 + per-request and per-row array."""

145 rng = np.random.default_rng(seed)

146 max_num_reqs = 64

147 batch = _make_batch(rng, num_reqs=23, max_num_reqs=max_num_reqs, max_spec=3)

@@ -225,8 +223,7 @@ def test_local_batch_partition(tp_size: int, seed: int):

223 def test_all_to_all_reassembly_simulated(tp_size: int):

224 """Vocab shards computed from owner-sorted hidden states and routed

225 through the simulated all-to-all reassemble into exactly the full-vocab

228 - logits of each rank's owned rows.

229 - """

226 + logits of each rank's owned rows."""

227 rng = np.random.default_rng(7)

228 max_num_reqs = 32

229 batch = _make_batch(rng, num_reqs=11, max_num_reqs=max_num_reqs, max_spec=4)

@@ -264,8 +261,7 @@ def test_all_to_all_reassembly_simulated(tp_size: int):

261 @requires_cuda

262 def test_shard_no_spec():

263 """Batches without draft tokens (one logits row per request) degenerate

267 - cleanly: rows mirror requests.

268 - """

264 + cleanly: rows mirror requests."""

265 rng = np.random.default_rng(11)

266 tp_size = 4

267 max_num_reqs = 32

@@ -284,8 +280,7 @@ def test_shard_no_spec():

280 @pytest.mark.parametrize("tp_size", [2, 4])

281 def test_shard_grammar_output(tp_size: int):

282 """The grammar filter keeps exactly the owned requests' bitmask rows, in

287 - grammar order, and the per-rank pieces partition the global bitmask.

288 - """

283 + grammar order, and the per-rank pieces partition the global bitmask."""

284 rng = np.random.default_rng(5)

285 max_num_reqs = 32

286 num_reqs = 12

@@ -340,8 +335,7 @@ def test_shard_grammar_output(tp_size: int):

335 @pytest.mark.parametrize("tp_size", [2, 4])

336 def test_gather_src_indices_round_trip(tp_size: int):

337 """Padded per-rank result blocks indexed by gathered_src_indices restore

343 - batch order, including when some rank owns zero requests.

344 - """

338 + batch order, including when some rank owns zero requests."""

339 rng = np.random.default_rng(3)

340 max_num_reqs = 40

341 all_slots = np.arange(max_num_reqs)

@@ -376,8 +370,7 @@ def test_gather_src_indices_round_trip(tp_size: int):

370 def test_gather_sampler_output_kernels(tp_size: int):

371 """The fused pack/unpack kernels round-trip per-rank sampler results

372 into batch order: token ids zero-filled past the local width, counts

379 - cast back to int32, identical on every rank.

380 - """

373 + cast back to int32, identical on every rank."""

374 from dataclasses import replace as dc_replace

375

376 from vllm.v1.worker.gpu.sample.output import SamplerOutput

@@ -472,8 +465,7 @@ def test_gather_sampler_output_kernels(tp_size: int):

465 def test_gather_sampler_output_logprobs_and_nans():

466 """num_nans is reduced per request and gathered; LogprobsTensors are

467 packed into fixed per-request blocks, width-aligned across ranks (pad

475 - and truncate paths), and restored to global row order.

476 - """

468 + and truncate paths), and restored to global row order."""

469 from dataclasses import replace as dc_replace

470

471 from vllm.v1.outputs import LogprobsTensors

@@ -606,8 +598,7 @@ def test_gather_sampler_output_logprobs_and_nans():

598 def test_finish_requests_frees_slots_in_sorted_order():

599 """Request slots must be freed in the same order on every TP rank:

600 ownership derives from slot indices, and `finished_req_ids` is a set

609 - whose iteration order is per-process hash-randomized.

610 - """

601 + whose iteration order is per-process hash-randomized."""

602 from vllm.v1.worker.gpu.model_runner import GPUModelRunner

603

604 removed: list[str] = []

tests/v1/worker/test_gpu_block_table.py

+2 -4

@@ -217,8 +217,7 @@ def test_v1_block_table_move_row_clears_vacated_row():

217 not keep stale block ids. Padded dummy-run batches dereference stale rows

218 as mamba state slots (bypassing the NULL_BLOCK_ID fill of real decode

219 padding) and write state in place there — corrupting the blocks' new

220 - owner once they are reallocated, e.g. to an in-flight NIXL load.

221 - """

220 + owner once they are reallocated, e.g. to an in-flight NIXL load."""

221 from vllm.v1.worker.block_table import BlockTable

222

223 block_table = BlockTable(

@@ -250,8 +249,7 @@ def test_get_dummy_block_tables_returns_zeroed_rows():

249 PAD-filled, state indices are not), so stale rows would direct dummy

250 state writes at freed — possibly reallocated — blocks.

251 get_dummy_block_tables must hand out zeroed (null block) rows while

253 - preserving the persistent storage address for CUDA graphs.

254 - """

252 + preserving the persistent storage address for CUDA graphs."""

253 device = torch.device("cuda")

254 block_tables = BlockTables(

255 block_sizes=[16],

tests/v1/worker/test_gpu_logit_bias.py

+1 -2

@@ -85,8 +85,7 @@ def test_v2_min_tokens_keeps_masking_when_other_candidates_remain():

85

86 def test_v2_min_tokens_mixed_batch_gates_restore_per_request():

87 """CHECK_ALL_MASKED_ROWS is a batch-wide constexpr, so a mixed batch must

88 - still consult the per-request flag.

89 - """

88 + still consult the per-request flag."""

89 out = _apply(_only_stop_token_left(2), structured=[True, False])

90

91 assert out[0, STOP_TOKEN] == 1.0

tests/v1/worker/test_gpu_model_runner.py

+2 -4

@@ -73,8 +73,7 @@ DEVICE_TYPE = current_platform.device_type

73 def _restore_default_dtype():

74 """Several tests here set the process-wide default dtype to float16 and

75 previously leaked it, corrupting later float-sensitive tests in the same

76 - pytest process (torch.randn silently produced fp16).

77 - """

76 + pytest process (torch.randn silently produced fp16)."""

77 old = torch.get_default_dtype()

78 yield

79 torch.set_default_dtype(old)

@@ -1454,8 +1453,7 @@ def test_v2_runner_snapshots_late_interleave_adjustment(monkeypatch):

1453

1454 def test_hybrid_block_table_initialization():

1455 """Test hybrid block table with different kernel and kvcache_manager block

1457 - sizes.

1458 - """

1456 + sizes."""

1457 from vllm.v1.worker.block_table import BlockTable

1458

1459 # Test configuration: kvcache_manager block size = 32,

tests/v1/worker/test_gpu_model_runner_mm_gather.py

+3 -6

@@ -62,8 +62,7 @@ def _gather(features, cached, *, num_scheduled, shift, num_computed=0):

62 def test_draft_shift_uses_boundary_feature_when_cached():

63 """The drafter's +1 look-ahead reaches the feature at offset ==

64 processed_end; when it is already cached it is used for the look-ahead

65 - position rather than ignored.

66 - """

65 + position rather than ignored."""

66 f0 = _feature("h0", offset=0, length=8)

67 f1 = _feature("h1", offset=8, length=8) # starts exactly at processed_end

68 mm_embeds, is_mm_embed = _gather([f0, f1], [f0, f1], num_scheduled=8, shift=1)

@@ -76,8 +75,7 @@ def test_draft_shift_uses_boundary_feature_when_cached():

75

76 def test_draft_shift_tolerates_missing_boundary_feature():

77 """When the +1 look-ahead feature past the processed boundary is not yet

79 - encoded, fall back to the token embedding instead of raising.

80 - """

78 + encoded, fall back to the token embedding instead of raising."""

79 f0 = _feature("h0", offset=0, length=8)

80 f1 = _feature("h1", offset=8, length=8) # boundary feature, not cached

81 mm_embeds, is_mm_embed = _gather([f0, f1], [f0], num_scheduled=8, shift=1)

@@ -89,8 +87,7 @@ def test_draft_shift_tolerates_missing_boundary_feature():

87

88 def test_draft_shift_raises_on_interior_miss():

89 """A miss for a feature within the processed range (not the look-ahead

92 - boundary) is a real invariant violation, even on the drafter path.

93 - """

90 + boundary) is a real invariant violation, even on the drafter path."""

91 f0 = _feature("h0", offset=0, length=8) # interior, within processed range

92 with pytest.raises(RuntimeError, match="Encoder cache miss"):

93 _gather([f0], [], num_scheduled=8, shift=1)

tests/v1/worker/test_gpu_model_runner_v2.py

+4 -8

@@ -219,8 +219,7 @@ def test_append_block_ids_rejects_write_past_row_capacity():

219

220 def _make_capture_runner(captured: bool) -> GPUModelRunner:

221 """Minimal V2 runner for capture_model: fakes everything except the

222 - cudagraph_manager's needs_capture decision.

223 - """

222 + cudagraph_manager's needs_capture decision."""

223 runner = GPUModelRunner.__new__(GPUModelRunner)

224 runner.model_state = SimpleNamespace(supports_mm_inputs=False)

225 runner.cudagraph_manager = SimpleNamespace(

@@ -246,8 +245,7 @@ def _make_capture_runner(captured: bool) -> GPUModelRunner:

245 def test_capture_model_locks_workspace_after_capture(monkeypatch):

246 """A workspace resize after capture frees the buffer the captured graphs

247 baked in, so capture_model must lock the workspace before returning

249 - (https://github.com/vllm-project/vllm/issues/55336).

250 - """

248 + (https://github.com/vllm-project/vllm/issues/55336)."""

249 runner = _make_capture_runner(captured=True)

250 monkeypatch.setattr(

251 model_runner_module, "freeze_gc_for_cudagraph_capture", contextlib.nullcontext

@@ -268,8 +266,7 @@ def test_capture_model_locks_workspace_after_capture(monkeypatch):

266

267 def test_capture_model_skips_lock_when_nothing_captured(monkeypatch):

268 """With no graphs to capture (e.g. enforce_eager) there is nothing baked

271 - into the workspace, so the early return must not lock it.

272 - """

269 + into the workspace, so the early return must not lock it."""

270 runner = _make_capture_runner(captured=False)

271 lock_calls = []

272 monkeypatch.setattr(

@@ -283,8 +280,7 @@ def test_capture_model_skips_lock_when_nothing_captured(monkeypatch):

280 def test_capture_model_profile_only_skips_lock(monkeypatch):

281 """The memory-profiling capture pass runs before kernel warmup and the

282 real capture; locking there would stop the warmup from growing the

286 - workspace to its scheduler-realistic size.

287 - """

283 + workspace to its scheduler-realistic size."""

284 runner = _make_capture_runner(captured=True)

285 monkeypatch.setattr(

286 model_runner_module, "freeze_gc_for_cudagraph_capture", contextlib.nullcontext

tests/v1/worker/test_gpu_model_runner_v2_cudagraph_profiling.py

+1 -2

@@ -82,8 +82,7 @@ def _make_profiling_runner(

82

83 class _FakePlatform:

84 """Stands in for current_platform; the global graph pool is a class attr,

85 - matching vllm.platforms.Platform's lazy singleton.

86 - """

85 + matching vllm.platforms.Platform's lazy singleton."""

86

87 _global_graph_pool: Any = GLOBAL_POOL

88

tests/v1/worker/test_gpu_profiler.py

+2 -4

@@ -120,8 +120,7 @@ def test_max_iterations(default_profiler_config):

120 def test_restart_after_max_iterations(default_profiler_config):

121 """A start_profile after an auto-stop must actually restart, not be

122 silently ignored (regression test: auto-stop used to leave _active

123 - True forever, so start() always bailed out early).

124 - """

123 + True forever, so start() always bailed out early)."""

124 default_profiler_config.max_iterations = 2

125 profiler = ConcreteWorkerProfiler(default_profiler_config)

126

@@ -332,8 +331,7 @@ class TestAnnotateProfile:

331

332 def test_profiler_entered_during_capture():

333 """Profiler is used as a context manager in _warmup_and_capture,

335 - confirming it is active during the actual graph capture run.

336 - """

334 + confirming it is active during the actual graph capture run."""

335 runner = MagicMock()

336 runner.compilation_config.cudagraph_num_of_warmups = 0

337 mock_profiler = MagicMock()

tests/v1/worker/test_gpu_rejection_sampler_i64.py

+1 -2

@@ -34,8 +34,7 @@ def _run(

34 ) -> tuple[torch.Tensor, torch.Tensor]:

35 """Run rejection_sample on a batch of uniform k=2 requests, with the

36 request under test at request index `roi_req` and draft-logits row

37 - `req_state_idx`. Returns that request's (sampled, num_sampled).

38 - """

37 + `req_state_idx`. Returns that request's (sampled, num_sampled)."""

38 num_logits = num_reqs * LOGITS_PER_REQ

39 roi_row = roi_req * LOGITS_PER_REQ

40

tests/v1/worker/test_gpu_thinking_budget.py

+1 -2

@@ -288,8 +288,7 @@ def test_v2_thinking_budget_clamps_oversized_budget():

288

289 def test_v2_thinking_budget_continues_end_prefix_from_prompt():

290 """A resumed prompt ending with a partial forced-end marker must not

291 - restart the marker sequence and duplicate its first token.

292 - """

291 + restart the marker sequence and duplicate its first token."""

292 req_states = _make_req_states([1, START, 10, 11, END_A], prompt_len=5)

293 state = ThinkingBudgetState(req_states, MockMultiTokenEndReasoningConfig())

294 state.add_request(3, SamplingParams(thinking_token_budget=3))

tests/v1/worker/test_gpu_ubatch_slicing.py

+1 -2

@@ -701,8 +701,7 @@ def test_microbatches_recompute_dcp_lens_from_truncated_seq_lens(dcp_rank: int):

701 @pytest.mark.skipif(not torch.cuda.is_available(), reason="triton kernel needs CUDA")

702 def test_each_microbatch_owns_its_dcp_buffer():

703 """Microbatch DCP lengths live in per-microbatch persistent buffers so

704 - CPU-visible shapes stay valid under CUDA graph capture and replay.

705 - """

704 + CPU-visible shapes stay valid under CUDA graph capture and replay."""

705 input_batch, _ = _make_cuda_input_batch([4, 4, 4, 4], [8, 9, 10, 11])

706

707 ubatches, dcp_buffers = _slice_with_dcp(input_batch, dcp_rank=0)

tests/v1/worker/test_gpu_worker.py

+5 -10

@@ -49,8 +49,7 @@ def plan_env(monkeypatch: pytest.MonkeyPatch, tmp_path):

49

50 def test_startup_plan_fingerprint_sensitivity(plan_env):

51 """The fingerprint is the OOM-safety key: stable for identical inputs,

52 - different for anything the profiled value depends on.

53 - """

52 + different for anything the profiled value depends on."""

53 fp = startup_plan.compute_plan_fingerprint

54 base = fp(_plan_worker().vllm_config, 0, 1)

55 assert base == fp(_plan_worker().vllm_config, 0, 1)

@@ -98,8 +97,7 @@ def _snapshot(free_memory, torch_memory=0):

97

98 def _profile_result(consumed, reserved_before=0, reserved_after=0):

99 """A result whose free-memory readings agree with `consumed`, which

101 - `memory_profiling` derives as the drop in free memory, negative when it grew.

102 - """

100 + `memory_profiling` derives as the drop in free memory, negative when it grew."""

101 return SimpleNamespace(

102 total_consumed=consumed,

103 transient_peak_headroom=0,

@@ -127,8 +125,7 @@ def test_profiling_fallback_declines_when_free_memory_dropped(rocm):

125 @pytest.mark.parametrize("rocm", [True], indirect=True)

126 def test_profiling_fallback_replaces_a_released_measurement(rocm):

127 """A negative measurement describes the rest of the device, so it is replaced

130 - by this process's reservation, which the rest of the device cannot move.

131 - """

128 + by this process's reservation, which the rest of the device cannot move."""

129 result = _profile_result(

130 consumed=-RELEASED_BY_OTHERS,

131 reserved_after=TORCH_RESERVED,

@@ -152,8 +149,7 @@ def test_profiling_fallback_never_returns_a_negative_amount(rocm):

149 @pytest.mark.parametrize("rocm", [False], indirect=True)

150 def test_profiling_fallback_declines_off_rocm(rocm):

151 """Platforms that account frees eagerly keep reporting the error, so the

155 - caller's assertion stays reachable there.

156 - """

152 + caller's assertion stays reachable there."""

153 result = _profile_result(consumed=-RELEASED_BY_OTHERS)

154

155 assert maybe_rocm_profiling_fallback(result) is None

@@ -177,8 +173,7 @@ def test_execute_model_waits_previous_pp_send_before_forward(

173 monkeypatch: pytest.MonkeyPatch,

174 ):

175 """Previous device handles are waited before the forward pass; the

180 - metadata handle is left to the GroupCoordinator's reaper.

181 - """

176 + metadata handle is left to the GroupCoordinator's reaper."""

177 import torch

178

179 from vllm.sequence import IntermediateTensors

tests/v1/worker/test_kv_block_zeroer.py

+1 -2

@@ -358,8 +358,7 @@ def test_warmup_respects_available_block_count():

358 def test_zeroes_exactly_one_block_per_layer(layout: KVCacheLayout):

359 """The zeroer must zero every byte of the target block in every layer and nothing

360 outside it — per head-group region under LHBNC, and never past the target block's

361 - tile under block-major layouts (no out-of-bounds writes, no clobbering).

362 - """

361 + tile under block-major layouts (no out-of-bounds writes, no clobbering)."""

362 device = torch.device("cuda")

363 num_blocks, num_layers = 4, 2

364 spec = FullAttentionSpec(

tests/v1/worker/test_mamba_utils.py

+4 -8

@@ -645,8 +645,7 @@ def test_mamba_groups_support_mixed_specs_in_uniform_group():

645

646 def _make_staging_ctx(max_num_reqs: int, device: torch.device) -> MagicMock:

647 """Build a MambaSpecDecodeGPUContext stand-in exposing only the four

648 - per-request staging buffers touched by stage_postprocess_inputs_to_gpu.

649 - """

648 + per-request staging buffers touched by stage_postprocess_inputs_to_gpu."""

649 ctx = MagicMock()

650 ctx.mamba_state_idx_buf = _MockCpuGpuBuffer(max_num_reqs, torch.int32, device)

651 ctx.num_scheduled_tokens_buf = _MockCpuGpuBuffer(max_num_reqs, torch.int32, device)

@@ -658,8 +657,7 @@ def _make_staging_ctx(max_num_reqs: int, device: torch.device) -> MagicMock:

657 def test_stage_postprocess_inputs_to_gpu_fills_pinned_views():

658 """stage_postprocess_inputs_to_gpu writes each request's state_idx and

659 scheduled/computed/draft counts into slot `i` for `req_ids[i]`, leaves

661 - slots past `num_reqs` untouched, and mirrors the values to the GPU buffers.

662 - """

660 + slots past `num_reqs` untouched, and mirrors the values to the GPU buffers."""

661 device = torch.device("cpu")

662 max_num_reqs = 8

663 ctx = _make_staging_ctx(max_num_reqs, device)

@@ -731,8 +729,7 @@ def test_stage_postprocess_inputs_to_gpu_fills_pinned_views():

729

730 def test_stage_postprocess_inputs_to_gpu_asserts_on_missing_state_idx():

731 """If preprocess_mamba didn't populate mamba_state_idx for a req in the

734 - batch, staging must fail loudly rather than silently writing a stale index.

735 - """

732 + batch, staging must fail loudly rather than silently writing a stale index."""

733 device = torch.device("cpu")

734 ctx = _make_staging_ctx(max_num_reqs=4, device=device)

735 scheduler_output = _make_postprocess_scheduler_output(

@@ -810,8 +807,7 @@ def _run_gpu_postprocess(

807 device: torch.device,

808 ) -> None:

809 """Initialize the GPU context against `block_table`, run the fused

813 - postprocess kernel for `req_ids`, and synchronize.

814 - """

810 + postprocess kernel for `req_ids`, and synchronize."""

811

812 def t(values):

813 return torch.tensor(values, dtype=torch.int32, device=device)

tests/v1/worker/test_mixed_warmup_gate.py

+1 -2

@@ -17,8 +17,7 @@ def _fail(*args, **kwargs):

17 @pytest.mark.parametrize("max_num_reqs", [1, 0])

18 def test_mixed_warmup_skipped_for_single_seq(max_num_reqs):

19 """A mixed prefill+decode step needs >=2 requests; with max_num_reqs < 2

20 - the warmup must be skipped without touching the worker callbacks.

21 - """

20 + the warmup must be skipped without touching the worker callbacks."""

21 runner = SimpleNamespace(is_pooling_model=False, max_num_reqs=max_num_reqs)

22

23 assert (

tests/v1/worker/test_prompt_embeds_state.py

+4 -8

@@ -78,8 +78,7 @@ def _apply(

78 def test_overlay_chunked_prefill_and_decode():

79 """Rows within the embeds range come from prompt_embeds at the

80 num_computed offset; requests without embeds and requests past their

81 - embeds length (decode) keep the base embedding.

82 - """

81 + embeds length (decode) keep the base embedding."""

82 state = _make_state()

83 embeds_a = torch.randn(6, HIDDEN, dtype=torch.float32)

84 embeds_b = torch.randn(5, HIDDEN, dtype=torch.float32)

@@ -98,8 +97,7 @@ def test_overlay_chunked_prefill_and_decode():

97

98 def test_overlay_clamps_to_embeds_length():

99 """A window straddling the end of the prompt embeds writes only the

101 - in-range rows (e.g. final prefill chunk + sampled token).

102 - """

100 + in-range rows (e.g. final prefill chunk + sampled token)."""

101 state = _make_state()

102 embeds = torch.randn(4, HIDDEN, dtype=torch.float32)

103 state.add_request(3, _NewReqData("a", embeds))

@@ -114,8 +112,7 @@ def test_overlay_clamps_to_embeds_length():

112

113 def test_overlay_respects_is_token_ids_mask():

114 """Mixed mode: positions marked as real token ids keep the base

117 - embedding; only embed positions are overwritten.

118 - """

115 + embedding; only embed positions are overwritten."""

116 state = _make_state()

117 embeds = torch.randn(5, HIDDEN, dtype=torch.float32)

118 is_token_ids = [True, False, False, True, False]

@@ -133,8 +130,7 @@ def test_overlay_respects_is_token_ids_mask():

130

131 def test_index_reuse_clears_stale_entry():

132 """A request added at a previously-used index without embeds must not

136 - inherit the prior occupant's pointer-table entry.

137 - """

133 + inherit the prior occupant's pointer-table entry."""

134 state = _make_state()

135 state.add_request(0, _NewReqData("a", torch.randn(4, HIDDEN, dtype=torch.float32)))

136 # A second live embeds request so the kernel actually launches (the

tests/v1/worker/test_workspace.py

+1 -2

@@ -84,8 +84,7 @@ def test_workspace_lanes_compose_with_ubatches(monkeypatch) -> None:

84

85 def test_workspace_lock_blocks_growth_and_unlock_restores(monkeypatch) -> None:

86 """Once locked, oversized requests fail loudly instead of reallocating the

87 - buffer that captured CUDA graphs point at; unlock restores growth.

88 - """

87 + buffer that captured CUDA graphs point at; unlock restores growth."""

88 monkeypatch.setattr(workspace, "dbo_current_ubatch_id", lambda: 0)

89 manager = workspace.WorkspaceManager(torch.device("cpu"), num_lanes=1)

90

tools/report_build_time_ninja.py

+3 -6

@@ -56,8 +56,7 @@ class Target:

56

57 def __init__(self, start, end):

58 """Creates a target object by passing in the start/end times in seconds

59 - as a float.

60 - """

59 + as a float."""

60 self.start = start

61 self.end = end

62 # A list of targets, appended to by the owner of this object.

@@ -105,8 +104,7 @@ class Target:

104 def ReadTargets(log, show_all):

105 """Reads all targets from .ninja_log file |log_file|, sorted by duration.

106

108 - The result is a list of Target objects.

109 - """

107 + The result is a list of Target objects."""

108 header = log.readline()

109 assert header == "# ninja log v5\n", "unrecognized ninja log version {!r}".format(

110 header

@@ -155,8 +153,7 @@ def GetExtension(target, extra_patterns):

153

154 For targets that generate multiple outputs it is important to return a

155 consistent 'canonical' extension. Ultimately the goal is to group build steps

158 - by type.

159 - """

156 + by type."""

157 for output in target.targets:

158 if extra_patterns:

159 for fn_pattern in extra_patterns.split(";"):

vllm/_aiter_ops.py

+1 -2

@@ -1974,8 +1974,7 @@ class rocm_aiter_ops:

1974 enabled aiter. Only aiter's Triton kernels exist on gfx12 (no CK build),

1975 so this deliberately stays off the gfx9/CK `@if_aiter_supported` umbrella

1976 and gates only the Triton paths rdna4 uses. The gfx12 analog of

1977 - `is_enabled()`.

1978 - """

1977 + `is_enabled()`."""

1978 if not current_platform.is_rocm() or not IS_AITER_FOUND:

1979 return False

1980 from vllm.platforms.rocm import on_rdna4

vllm/benchmarks/datasets/create_txt_slices_dataset.py

+1 -2

@@ -65,8 +65,7 @@ def create_txt_slices_jsonl(

65 trust_remote_code: bool = False,

66 ) -> None:

67 """Read *input_path*, slice it into prompts, and write JSONL to

68 - *output_path*.

69 - """

68 + *output_path*."""

69 tokenizer = AutoTokenizer.from_pretrained(

70 tokenizer_name, trust_remote_code=trust_remote_code

71 )

vllm/benchmarks/datasets/datasets.py

+1 -2

@@ -4383,8 +4383,7 @@ class PrefixRepetitionRandomDataset(BenchmarkDataset):

4383

4384 def _generate_exact_length_tokens(target_length: int) -> tuple[list[int], int]:

4385 """Generate tokens that decode and re-encode to exactly

4386 - target_length.

4387 - """

4386 + target_length."""

4387 # Generate random tokens

4388 tokens = np.random.randint(0, vocab_size, size=target_length).tolist()

4389

vllm/benchmarks/lib/endpoint_request_func.py

+2 -4

@@ -22,8 +22,7 @@ AIOHTTP_TIMEOUT = aiohttp.ClientTimeout(total=6 * 60 * 60)

22

23 class StreamedResponseHandler:

24 """Handles streaming HTTP responses by accumulating chunks until complete

25 - messages are available.

26 - """

25 + messages are available."""

26

27 def __init__(self):

28 self.buffer = ""

@@ -31,8 +30,7 @@ class StreamedResponseHandler:

30

31 def add_chunk(self, chunk_bytes: bytes) -> list[str]:

32 """Add a chunk of bytes to the buffer and return any complete

34 - messages.

35 - """

33 + messages."""

34 chunk_str = self._decoder.decode(chunk_bytes)

35 self.buffer += chunk_str

36

vllm/compilation/backends.py

+1 -2

@@ -150,8 +150,7 @@ class CompilerManager:

150 def compile_context(self, compile_range: Range) -> Generator[None, None, None]:

151 """Provide compilation context for the duration of compilation to set

152 any torch global properties we want to scope to a single Inductor

153 - compilation (e.g. partition rules, pass context).

154 - """

153 + compilation (e.g. partition rules, pass context)."""

154 with pass_context(compile_range):

155 if self.compilation_config.use_inductor_graph_partition:

156 with inductor_partition_rule_context(

vllm/compilation/compiler_interface.py

+1 -2

@@ -153,8 +153,7 @@ def _get_vllm_functorch_config() -> dict[str, Any]:

153 """Return the functorch config overrides that vLLM applies at compile time.

154

155 Used by both set_functorch_config() and get_inductor_factors() to ensure

156 - the compile-time config and cache key are always consistent.

157 - """

156 + the compile-time config and cache key are always consistent."""

157 cfg: dict[str, Any] = {}

158 if not envs.VLLM_USE_MEGA_AOT_ARTIFACT:

159 cfg["bundled_autograd_cache"] = False

vllm/compilation/passes/fusion/allreduce_rms_fusion.py

+1 -2

@@ -66,8 +66,7 @@ def _view_flashinfer_nvfp4_scale_out_as_int32(

66

67 def _norm_input_weight_dtype_match(match: pm.Match) -> bool:

68 """Prevent fusion when the norm input and weight dtypes differ (e.g. a Gemma

69 - fp32 weight.float()+1 gamma), covering rms_norm and fused_add_rms_norm.

70 - """

69 + fp32 weight.float()+1 gamma), covering rms_norm and fused_add_rms_norm."""

70 for node in match.nodes:

71 if node.target == _IR_RMS_NORM_OP:

72 x, weight = node.args[0], node.args[1]

vllm/compilation/passes/utility/split_coalescing.py

+1 -2

@@ -30,8 +30,7 @@ logger = init_logger(__name__)

30

31 class SplitCoalescingPass(VllmInductorPass):

32 """Replace duplicate ``split_with_sizes`` nodes with a single canonical

33 - node when they share the same input tensor and split sizes.

34 - """

33 + node when they share the same input tensor and split sizes."""

34

35 @VllmInductorPass.time_and_log

36 def __call__(self, graph: fx.Graph) -> None:

vllm/config/compilation.py

+2 -4

@@ -36,8 +36,7 @@ logger = init_logger(__name__)

36

37 class CompilationMode(enum.IntEnum):

38 """The compilation approach used for torch.compile-based compilation of the

39 - model.

40 - """

39 + model."""

40

41 NONE = 0

42 """No torch.compile compilation is applied, model runs in fully eager pytorch mode.

@@ -337,8 +336,7 @@ class PassConfig:

336 class DynamicShapesType(str, enum.Enum):

337 """Types of dynamic shapes handling in torch.compile().

338 see Dynamic shapes and vllm guard dropping in torch_compile.md

340 - for more details.

341 - """

339 + for more details."""

340

341 BACKED = "backed"

342 """Use backed dynamic shapes. torch.compile() guards on backed dynamic

vllm/config/model.py

+1 -2

@@ -1005,8 +1005,7 @@ class ModelConfig:

1005

1006 def _get_transformers_backend_cls(self) -> str:

1007 """Determine which Transformers modeling backend class will be used if

1008 - `model_impl` is set to `transformers` or `auto`.

1009 - """

1008 + `model_impl` is set to `transformers` or `auto`."""

1009 cls = "Transformers"

1010 # If 'hf_config is not hf_text_config' it's a nested config, i.e. multimodal

1011 cls += "MultiModal" if self.hf_config is not self.hf_text_config else ""

vllm/config/multimodal.py

+1 -2

@@ -524,8 +524,7 @@ class MultiModalConfig:

524

525 def get_video_pruning_spec(self) -> tuple[VideoPruningMethod, float] | None:

526 """Return `(method, rate)` when video pruning is enabled, else None.

527 - `rate` is the fraction of video tokens to prune.

528 - """

527 + `rate` is the fraction of video tokens to prune."""

528 if self.video_pruning_rate is not None and self.video_pruning_rate > 0:

529 return (self.video_pruning_method, float(self.video_pruning_rate))

530 return None

vllm/config/observability.py

+1 -2

@@ -156,8 +156,7 @@ class ObservabilityConfig:

156 cls, value: list[DetailedTraceModules] | None

157 ) -> list[DetailedTraceModules] | None:

158 """Handle the legacy case where users might provide a comma-separated

159 - string instead of a list of strings.

160 - """

159 + string instead of a list of strings."""

160 if value is not None and len(value) == 1 and "," in value[0]:

161 value = cast(list[DetailedTraceModules], value[0].split(","))

162 return value

vllm/config/reasoning.py

+2 -4

@@ -47,15 +47,13 @@ class ReasoningConfig:

47 @property

48 def enabled(self) -> bool:

49 """Returns True if reasoning is enabled (i.e. if token IDs have been

50 - initialized), False otherwise.

51 - """

50 + initialized), False otherwise."""

51 return self._enabled

52

53 @property

54 def reasoning_start_token_ids(self) -> list[int] | None:

55 """Token IDs derived from `reasoning_start_str`. Set automatically by

57 - `initialize_token_ids`. Not intended to be configured directly.

58 - """

56 + `initialize_token_ids`. Not intended to be configured directly."""

57 return self._reasoning_start_token_ids

58

59 @property

vllm/config/speculative.py

+2 -4

@@ -543,8 +543,7 @@ class SpeculativeConfig:

543 @staticmethod

544 def _acceptance_length_to_rates(length: float, n: int) -> list[float]:

545 """Mean acceptance length to unconditional per-position rates, using

546 - the minimum-variance schedule.

547 - """

546 + the minimum-variance schedule."""

547 num_drafts = length - 1 # expected number of accepted draft tokens

548 num_full = int(num_drafts)

549 return (

@@ -558,8 +557,7 @@ class SpeculativeConfig:

557 length: float | None,

558 ) -> list[float]:

559 """Return per-position unconditional acceptance rates from exactly one

561 - of `rates` or `length` (validates range, length, and monotonicity).

562 - """

560 + of `rates` or `length` (validates range, length, and monotonicity)."""

561 if (rates is None) == (length is None):

562 raise ValueError(

563 "rejection_sample_method='synthetic' requires exactly one of "

vllm/config/utils.py

+2 -4

@@ -84,8 +84,7 @@ def config(

84

85 def get_field(cls: ConfigType, name: str) -> Any:

86 """Get the default factory field of a dataclass by name. Used for getting

87 - default factory fields in `EngineArgs`.

88 - """

87 + default factory fields in `EngineArgs`."""

88 if not is_dataclass(cls):

89 raise TypeError("The given class is not a dataclass.")

90 try:

@@ -122,8 +121,7 @@ def is_init_field(cls: ConfigType, name: str) -> bool:

121 def replace(dataclass_instance: ConfigT, /, **kwargs) -> ConfigT:

122 """Like [`dataclasses.replace`](https://docs.python.org/3/library/dataclasses.html#dataclasses.replace),

123 but compatible with Pydantic dataclasses which use `pydantic.fields.Field` instead

125 - of `dataclasses.field`

126 - """

124 + of `dataclasses.field`"""

125 cls = type(dataclass_instance)

126 dataclass_dict = dataclass_instance.__dict__

127 dataclass_dict = {k: v for k, v in dataclass_dict.items() if is_init_field(cls, k)}

vllm/config/vllm.py

+2 -4

@@ -153,8 +153,7 @@ IS_DENSE = False

153

154 def enable_norm_fusion(cfg: "VllmConfig") -> bool:

155 """Enable if either RMS norm or quant FP8 custom op is active;

156 - otherwise Inductor handles fusion.

157 - """

156 + otherwise Inductor handles fusion."""

157 return (

158 cfg.compilation_config.is_custom_op_enabled("rms_norm")

159 or cfg.compilation_config.is_custom_op_enabled("quant_fp8")

@@ -744,8 +743,7 @@ class VllmConfig:

743

744 def _is_dflash2_draft(self) -> bool:

745 """Whether the DFlash draft is a DFlash2 one, by the architecture the

747 - speculator selects on (v1/worker/gpu/spec_decode/__init__.py).

748 - """

746 + speculator selects on (v1/worker/gpu/spec_decode/__init__.py)."""

747 spec = self.speculative_config

748 if spec is None or spec.method != "dflash":

749 return False

vllm/device_allocator/sleep_mode_backend.py

+5 -10

@@ -70,8 +70,7 @@ class SleepModeBackend(ABC):

70

71 def state(self) -> SleepModeState:

72 """Current lifecycle state. Lets ``/health`` distinguish a healthy-idle

73 - (suspended) engine from a healthy-serving one (see RFC #34303).

74 - """

73 + (suspended) engine from a healthy-serving one (see RFC #34303)."""

74 return self._state

75

76 # -- Capability introspection (no instance required) --

@@ -84,30 +83,26 @@ class SleepModeBackend(ABC):

83 @classmethod

84 def preserves_communicators(cls) -> bool:

85 """If False, collective communicators (e.g. NCCL) are destroyed by

87 - ``suspend`` and the executor must re-initialize them on ``resume``.

88 - """

86 + ``suspend`` and the executor must re-initialize them on ``resume``."""

87 return False

88

89 @classmethod

90 def preserves_compiled_artifacts(cls) -> bool:

91 """If True, torch.compile / JIT kernels survive suspend/resume and need

94 - not be recompiled on resume.

95 - """

92 + not be recompiled on resume."""

93 return False

94

95 @classmethod

96 def preserves_graphs_with_communicators(cls) -> bool:

97 """If True, CUDA graphs containing collective communicators (e.g. NCCL)

98 stay valid after resume. False when communicators are rebuilt (embedded

102 - comm handles go stale).

103 - """

99 + comm handles go stale)."""

100 return False

101

102 @classmethod

103 def supports_durable_storage(cls) -> bool:

104 """If True, suspended state can be persisted beyond the process

109 - lifetime (disk or object storage) and restored in a new process.

110 - """

105 + lifetime (disk or object storage) and restored in a new process."""

106 return False

107

108

vllm/distributed/device_communicators/all_reduce_utils.py

+1 -2

@@ -156,8 +156,7 @@ def should_nccl_symm_mem_allreduce(world_size: int, input_tensor: torch.Tensor)

156

157 def should_nccl_symm_mem_ag_rs() -> bool:

158 """Check whether NCCL symmetric memory should be used for

159 - AllGather / ReduceScatter collectives.

160 - """

159 + AllGather / ReduceScatter collectives."""

160 from vllm.distributed.device_communicators.pynccl_allocator import (

161 is_symmetric_memory_enabled,

162 )

vllm/distributed/device_communicators/shm_broadcast.py

+2 -4

@@ -650,8 +650,7 @@ class MessageQueue:

650

651 def shutdown(self):

652 """If this is an idle reader, wakes it up so it can clean up and shut

653 - down

654 - """

653 + down"""

654 self.shutting_down = True

655 if self._spin_condition is not None:

656 self._spin_condition.cancel()

@@ -761,8 +760,7 @@ class MessageQueue:

760

761 def should_warn(self) -> bool:

762 """Returns true if it's time to log a warning for a timeout that is not

764 - indefinite

765 - """

763 + indefinite"""

764 if self._should_warn:

765 elapsed = time.monotonic() - self.started

766 if elapsed >= VLLM_RINGBUFFER_WARNING_INTERVAL * self.n_warning:

vllm/distributed/ec_transfer/ec_connector/cpu/scheduler/embedding_cache.py

+1 -2

@@ -15,8 +15,7 @@ from collections import OrderedDict

15

16 class CacheEntry:

17 """A single cache entry. Read `.block_ids` and `.ready` freely;

18 - mutations only through EmbeddingCache methods.

19 - """

18 + mutations only through EmbeddingCache methods."""

19

20 __slots__ = ("block_ids", "_pin_count")

21

vllm/distributed/eplb/eplb_communicator.py

+2 -4

@@ -84,8 +84,7 @@ class EplbCommunicator(ABC):

84 @property

85 def needs_profile_buffer_reservation(self) -> bool:

86 """Whether the profile path must run a dummy collective operation to reserve

87 - communication buffers.

88 - """

87 + communication buffers."""

88 return True

89

90 def set_stream(self, cuda_stream: torch.cuda.Stream | None) -> None:

@@ -443,8 +442,7 @@ class NixlEplbCommunicator(EplbCommunicator):

442

443 def _exchange_remote_send_meta(self) -> None:

444 """Exchange per-layer per-tensor metadata so receivers can compute

446 - remote RDMA addresses at transfer time.

447 - """

445 + remote RDMA addresses at transfer time."""

446 local_meta: dict[tuple[int, int], tuple[int, int, int]] = {}

447 for layer_idx, layer_tensors in enumerate(self._all_expert_weights):

448 for t_idx, t in enumerate(layer_tensors):

vllm/distributed/kv_transfer/kv_connector/utils.py

+1 -2

@@ -51,8 +51,7 @@ def get_kv_connector_cache_layout(vllm_config: VllmConfig | None = None):

51

52 class KVOutputAggregator:

53 """Utility class to aggregate the output of all workers into a single

54 - output corresponding to Rank 0 for scheduler.

55 - """

54 + output corresponding to Rank 0 for scheduler."""

55

56 def __init__(self, expected_finished_count: int):

57 # Complete transfer tracker. Used to track finished requests

vllm/distributed/kv_transfer/kv_connector/v1/example_hidden_states_connector.py

+1 -2

@@ -278,8 +278,7 @@ class ExampleHiddenStatesConnector(KVConnectorBase_V1, SupportsHMA):

278

279 def _on_write_done(self, req_id: str, future: Future) -> None:

280 """Surface any exception from the disk-write thread and drop the

281 - completed future from the in-flight tracking dict.

282 - """

281 + completed future from the in-flight tracking dict."""

282 self._req_futures.pop(req_id, None)

283 exc = future.exception()

284 if exc is not None:

vllm/distributed/kv_transfer/kv_connector/v1/lmcache_mp_connector.py

+2 -4

@@ -231,8 +231,7 @@ class LMCacheMPRequestTracker:

231 ####

232 def needs_retrieve(self) -> bool:

233 """Check whether the current request needs retrieve, will be used

234 - update_stage_after_alloc

235 - """

234 + update_stage_after_alloc"""

235 return (

236 self.num_lmcache_hit_blocks > self.num_vllm_hit_blocks

237 and self.state != LMCacheMPRequestState.READY

@@ -240,8 +239,7 @@ class LMCacheMPRequestTracker:

239

240 def is_ready_for_retrieving(self) -> bool:

241 """Check whether the current request is ready for retrieving,

243 - will be used in process_loading_requests

244 - """

242 + will be used in process_loading_requests"""

243 return (

244 self.state == LMCacheMPRequestState.WAITING_FOR_LOAD

245 and self.needs_retrieve()

vllm/distributed/kv_transfer/kv_connector/v1/mooncake/mooncake_connector.py

+4 -8

@@ -718,8 +718,7 @@ class MooncakeConnectorScheduler:

718

719 def _get_remote_prefill_token_count(self, num_prompt_tokens: int) -> int:

720 """D-side only. Returns N-1 for Mamba models since the decoder

721 - always recomputes the last token and must start from h(N-1).

722 - """

721 + always recomputes the last token and must start from h(N-1)."""

722 if self._has_mamba and num_prompt_tokens > 1:

723 return num_prompt_tokens - 1

724 return num_prompt_tokens

@@ -730,8 +729,7 @@ class MooncakeConnectorScheduler:

729 derive h(N) correctly.

730

731 Guarded by ``_p_side_truncated`` to avoid repeated truncation if the

733 - request is preempted and rescheduled.

734 - """

732 + request is preempted and rescheduled."""

733 params = request.kv_transfer_params

734 if (

735 params is not None

@@ -1677,8 +1675,7 @@ class MooncakeConnectorWorker:

1675 def _bind_sender_thread_device(self) -> None:

1676 """ThreadPoolExecutor initializer — binds each pool thread to the

1677 correct CUDA device. CUDA device selection is thread-local, so

1680 - without this, NVLink transfers fail for TP ranks > 0.

1681 - """

1678 + without this, NVLink transfers fail for TP ranks > 0."""

1679 current_platform.set_device(self.device_id)

1680

1681 def _send_blocks(

@@ -1904,8 +1901,7 @@ class MooncakeConnectorWorker:

1901

1902 def get_kv_connector_stats(self) -> KVConnectorStats | None:

1903 """Return transfer stats collected since the last call, or None

1907 - if nothing has been recorded in this interval.

1908 - """

1904 + if nothing has been recorded in this interval."""

1905 if self.xfer_stats.is_empty():

1906 return None

1907 return self.xfer_stats.clone_and_reset()

vllm/distributed/kv_transfer/kv_connector/v1/mooncake/store/coordinator.py

+1 -2

@@ -73,8 +73,7 @@ class ExternalCachedBlockPool:

73

74 class MooncakeStoreCoordinator:

75 """Mirror of ``HybridKVCacheCoordinator.find_longest_cache_hit`` over an

76 - ``ExternalCachedBlockPool``.

77 - """

76 + ``ExternalCachedBlockPool``."""

77

78 def __init__(

79 self,

vllm/distributed/kv_transfer/kv_connector/v1/mooncake/store/worker.py

+1 -2

@@ -2629,8 +2629,7 @@ class LookupKeyClient:

2629 non_block: bool = False,

2630 ) -> MooncakeLookupResult | None:

2631 """If non_block is True, will return None until the result is ready,

2632 - so the caller retries on a later step.

2633 - """

2632 + so the caller retries on a later step."""

2633 future = self.futures.get(req_id)

2634 if future is None:

2635 future = self.executor.submit(self._lookup, num_tokens, list(block_hashes))

vllm/distributed/kv_transfer/kv_connector/v1/moriio/moriio_connector.py

+1 -2

@@ -548,8 +548,7 @@ class MoRIIOConnectorScheduler:

548 self, request_id: ReqId, transfer_id: TransferId | None = None

549 ):

550 """Unmap request_id/transfer_id. Uses transfer_id for lookup if

551 - exact request_id match fails (handles input_processor mutation).

552 - """

551 + exact request_id match fails (handles input_processor mutation)."""

552 if request_id in self.request_id_to_transfer_id:

553 tid = self.request_id_to_transfer_id[request_id]

554 del self.request_id_to_transfer_id[request_id]

vllm/distributed/kv_transfer/kv_connector/v1/nixl/base_scheduler.py

+2 -4

@@ -400,8 +400,7 @@ class NixlBaseConnectorScheduler:

400 def _get_remote_prefill_token_count(self, num_prompt_tokens: int) -> int:

401 """D-side only. The number of prompt tokens to load from the prefiller.

402 Stops short of the trailing ``_prefill_backoff()`` tokens that the decoder

403 - will recompute locally.

404 - """

403 + will recompute locally."""

404 backoff = self._prefill_backoff()

405 if backoff and num_prompt_tokens > backoff:

406 return num_prompt_tokens - backoff

@@ -414,8 +413,7 @@ class NixlBaseConnectorScheduler:

413 multi-module MTP it is the drafter's whole lookahead window.

414

415 Guarded by ``_p_side_truncated`` to avoid repeated truncation if the

417 - request is preempted and rescheduled.

418 - """

416 + request is preempted and rescheduled."""

417 backoff = self._prefill_backoff()

418 params = request.kv_transfer_params

419 if (

vllm/distributed/kv_transfer/kv_connector/v1/nixl/base_worker.py

+2 -4

@@ -1873,8 +1873,7 @@ class NixlBaseConnectorWorker:

1873 ) -> np.ndarray:

1874 """Build remote desc regions (conv sub-projections + ssm) per layer.

1875 For hetero-TP, each D rank reads only its sub-projection slice from

1876 - the P rank. Returns an Nx3 uint64 array.

1877 - """

1876 + the P rank. Returns an Nx3 uint64 array."""

1877 assert nixl_agent_meta.kv_caches_base_addr, (

1878 "Remote KV cache base addresses must not be empty."

1879 )

@@ -2657,8 +2656,7 @@ class NixlBaseConnectorWorker:

2656 @cached_property

2657 def _attention_kv_caches(self) -> list[torch.Tensor]:

2658 """Device KV caches of attention layers (mamba states excluded),

2660 - as consumed by the receive post-process.

2661 - """

2659 + as consumed by the receive post-process."""

2660 assert self.device_kv_caches, (

2661 "_attention_kv_caches accessed before register_kv_caches"

2662 )

vllm/distributed/kv_transfer/kv_connector/v1/nixl/push_worker.py

+2 -4

@@ -315,8 +315,7 @@ class NixlPushConnectorWorker(NixlBaseConnectorWorker):

315 executes on the writer; the handshake runs on the background executor

316 and the request is re-queued onto ``_reg_send_inbox`` once it

317 completes (at which point ``_ensure_handshake`` returns ``None`` and we

318 - send directly).

319 - """

318 + send directly)."""

319 remote_pp_size = reg_data.get("remote_pp_size", 1)

320 fut = self._ensure_handshake(

321 reg_data["remote_engine_id"],

@@ -508,8 +507,7 @@ class NixlPushConnectorWorker(NixlBaseConnectorWorker):

507 ``BlockIds`` is canonically a tuple of per-group lists, but some

508 registration payloads collapse a single-group case to a flat

509 list. Re-wrap that case so downstream group-aware helpers see a

511 - consistent shape.

512 - """

510 + consistent shape."""

511 if block_ids and not isinstance(block_ids[0], (list, tuple)):

512 return (list(block_ids),)

513 return block_ids

vllm/distributed/kv_transfer/kv_connector/v1/offloading/canonical_mapping.py

+4 -8

@@ -37,8 +37,7 @@ def canonical_format_id(kv_cache_layout: str) -> str:

37 Canonical pages keep the worker's KV layout family, so the id couples the

38 format version with that family; consumers must match it exactly. The family

39 keeps its historical NHD/HND spelling so ids stay stable for KV persisted

40 - before the layout enum existed.

41 - """

40 + before the layout enum existed."""

41 layout = KVCacheLayout[kv_cache_layout]

42 legacy = {KVCacheLayout.LBNHC: "nhd", KVCacheLayout.LBHNC: "hnd"}

43 family = legacy.get(layout, layout.name.lower())

@@ -122,8 +121,7 @@ def _interleave_cp_tokens(

121 ctx: _RankContext,

122 ) -> tuple[CopyRun, ...]:

123 """Place each region's num_tokens rows at their canonical token positions,

125 - one run per chunk of interleaved tokens.

126 - """

124 + one run per chunk of interleaved tokens."""

125 runs: list[CopyRun] = []

126 for region in regions:

127 if ctx.cp_size == 1:

@@ -258,8 +256,7 @@ def _attention_byte_regions(

256 cp_size: int,

257 ) -> list[ByteRegion] | None:

258 """Byte regions of an attention page, given this rank's head shard.

261 - None when the physical layout is not recognized (fail closed).

262 - """

259 + None when the physical layout is not recognized (fail closed)."""

260 bs, heads, head_size = spec.block_size, spec.num_kv_heads, spec.head_size

261 if tuple(kv_cache.shape) == (num_blocks, heads, bs, 2 * head_size):

262 return _packed_kv_regions(kv_cache, spec, head_shard, num_head_shards, cp_size)

@@ -373,8 +370,7 @@ def _is_exact_partition(intervals: list[tuple[int, int]], size: int) -> bool:

370

371 def _verify_tiling(layer_name: str, per_rank: list[CanonicalPageMapping]) -> None:

372 """Whichever ranks a block elects as writers must tile the canonical page

376 - exactly once, and each rank's runs must cover exactly its local page.

377 - """

373 + exactly once, and each rank's runs must cover exactly its local page."""

374 size = per_rank[0].canonical_page_size_bytes

375 num_writers = per_rank[0].num_writers

376 for mapping in per_rank:

vllm/distributed/kv_transfer/kv_connector/v1/offloading/events.py

+3 -6

@@ -79,8 +79,7 @@ def get_offloading_event_group_spec(

79 class _OffloadEventMetadata:

80 """BlockStored payload snapshot for one OffloadKey, captured while the

81 Request is available and kept until the final matching removal event.

82 - ``medium`` and ``ownership`` are forwarded from the OffloadingEvent.

83 - """

82 + ``medium`` and ``ownership`` are forwarded from the OffloadingEvent."""

83

84 # The chunk's constituent block hashes; the last one is the OffloadKey.

85 block_hashes: tuple[BlockHash, ...]

@@ -256,8 +255,7 @@ class OffloadingEventsTracker:

255

256 def reset(self) -> None:

257 """Drop all tracked state; pending payloads are stale after a

259 - manager cache reset.

260 - """

258 + manager cache reset."""

259 self._pending_event_metadata.clear()

260

261 def _build_event_metadata(

@@ -268,8 +266,7 @@ class OffloadingEventsTracker:

266 ) -> _OffloadEventMetadata:

267 """Build the payload snapshot for one offloaded chunk: its

268 constituent per-block hashes, the whole chunk's tokens, and the

271 - per-block ``block_size``.

272 - """

269 + per-block ``block_size``."""

270 hashes_per_chunk = group_config.hashes_per_chunk

271 assert hashes_per_chunk > 0

272 assert chunk_idx >= 0

vllm/distributed/kv_transfer/kv_connector/v1/offloading/scheduler.py

+2 -4

@@ -676,8 +676,7 @@ class OffloadingConnectorScheduler:

676 start_chunk_idx: int,

677 ) -> int | None:

678 """Return the number of consecutive offloaded chunks from the start,

679 - or None if the backend deferred a lookup.

680 - """

679 + or None if the backend deferred a lookup."""

680 hit_count = 0

681 defer_lookup = False

682 for local_idx, key in enumerate(keys):

@@ -712,8 +711,7 @@ class OffloadingConnectorScheduler:

711 """Return the end index (in `keys`) of the last run of

712 `sliding_window_size` consecutive hits, scanning from the end.

713 The first run may need a larger window for a partial rightmost chunk.

715 - Returns 0 on miss, None if the backend deferred a lookup.

716 - """

714 + Returns 0 on miss, None if the backend deferred a lookup."""

715 defer_lookup = False

716 pending_in_window = False

717 consecutive_hits = 0

vllm/distributed/utils.py

+1 -2

@@ -59,8 +59,7 @@ def ensure_divisibility(numerator, denominator):

59

60 def divide(numerator, denominator):

61 """Ensure that numerator is divisible by the denominator and return

62 - the division value.

63 - """

62 + the division value."""

63 ensure_divisibility(numerator, denominator)

64 return numerator // denominator

65

vllm/distributed/weight_transfer/factory.py

+1 -2

@@ -141,8 +141,7 @@ class WeightTransferTrainerFactory:

141 class_name: str | None = None,

142 ) -> None:

143 """Register a trainer engine. Same conventions as

144 - `WeightTransferEngineFactory.register_engine`.

145 - """

144 + `WeightTransferEngineFactory.register_engine`."""

145 if name in cls._registry:

146 raise ValueError(

147 f"Weight transfer trainer engine '{name}' is already registered."

vllm/distributed/weight_transfer/ipc_engine.py

+2 -4

@@ -40,8 +40,7 @@ class IPCWeightTransferInitInfo(WeightTransferInitInfo):

40 `packed` is a must-agree wire param: the trainer ships it here at the init

41 handshake so the worker decodes with the same setting the trainer encoded

42 with. The consumer rebuilds from the IPC handle + `tensor_sizes`, so it does

43 - not need the buffer size (producer-only).

44 - """

43 + not need the buffer size (producer-only)."""

44

45 packed: bool = False

46

@@ -54,8 +53,7 @@ class IPCTrainerInitInfo(TrainerInitInfo):

53

54 `packed` / `packed_buffer_size_bytes` are the transfer's wire params. The

55 trainer propagates `packed` to the worker at `trainer_init` so the two sides

57 - cannot disagree. `backend` is the factory dispatch key.

58 - """

56 + cannot disagree. `backend` is the factory dispatch key."""

57

58 backend: ClassVar[str] = "ipc"

59 packed: bool = False

vllm/distributed/weight_transfer/nccl_common.py

+1 -2

@@ -118,8 +118,7 @@ def worker_init_payload(init_info: NCCLWeightTransferInitInfo) -> dict:

118 """Serialize a worker init info for `init_weight_transfer_engine`, dropping

119 the unset rendezvous field (the UID in TCP mode) so the wire payload carries

120 only the mode actually in use. Shared by the dense and sparse trainer

121 - engines so the two cannot drift.

122 - """

121 + engines so the two cannot drift."""

122 return {key: value for key, value in asdict(init_info).items() if value is not None}

123

124

vllm/distributed/weight_transfer/nccl_engine.py

+2 -4

@@ -69,8 +69,7 @@ class NCCLTrainerInitInfo(TrainerInitInfo):

69 propagates them to the worker at `trainer_init` so the two sides cannot

70 disagree. Note this defaults to packed, unlike the worker-side

71 `NCCLWeightTransferInitInfo`, whose default only applies when no trainer

72 - ships a value. `backend` is the factory dispatch key.

73 - """

72 + ships a value. `backend` is the factory dispatch key."""

73

74 backend: ClassVar[str] = "nccl"

75

@@ -365,8 +364,7 @@ class NCCLTrainerWeightTransferEngine(TrainerWeightTransferEngine[NCCLTrainerIni

364 def _broadcast(self, source: WeightSource, meta: list[ParamMeta]) -> None:

365 """Iterate the source (materializing each tensor — a collective on all

366 ranks) and, on the sender, broadcast from rank 0, packed or one-by-one.

368 - Non-sender ranks only replay the iteration to stay in the collective.

369 - """

367 + Non-sender ranks only replay the iteration to stay in the collective."""

368 if not self.is_sender:

369 for _ in source:

370 pass

vllm/distributed/weight_transfer/sharded_rdt_common.py

+3 -6

@@ -214,8 +214,7 @@ class RdtRouter:

214

215 def class_of(self, name: str) -> int:

216 """The name's owner class — the planner's bucketing key, since all names

217 - of a chunk must share one producer.

218 - """

217 + of a chunk must share one producer."""

218 return self._class_of.get(name, 0)

219

220 def owners(self, name: str) -> list[int]:

@@ -286,8 +285,7 @@ class RdtRouter:

285 def bind(self, actors: list, produce_methods: list, consumer_id: int) -> None:

286 """Attach this consumer's producer handles. Rebuilt wholesale per init,

287 never appended to: every owner index is a position in these lists, so a

289 - rejoining engine that re-inits must not shift them.

290 - """

288 + rejoining engine that re-inits must not shift them."""

289 if len(actors) != len(produce_methods):

290 raise ValueError("actors and produce_methods must be parallel")

291 self._actors = list(actors)

@@ -350,6 +348,5 @@ def buffer_alloc_bytes(nbytes: int, presize: int = 0) -> int:

348 optional ``presize`` floor, and a coarse 256MB round-up, so the buffer is

349 allocated ONCE and never regrows. Regrowth is a correctness hazard, not just a

350 perf one -- see ``buffer_presize_gb``. Shared by the consumer's receive buffers

353 - and the producer's serve rings.

354 - """

351 + and the producer's serve rings."""

352 return max(nbytes, presize, -(-nbytes // (256 << 20)) * (256 << 20))

vllm/distributed/weight_transfer/sharded_rdt_engine.py

+16 -32

@@ -134,8 +134,7 @@ class _PendingPull:

134

135 ``targets``/``blob`` must stay strongly referenced until it completes:

136 ``set_target_for_ref`` stores WEAKREFS, so dropping them silently reroutes the

137 - transfer into a fallback buffer.

138 - """

137 + transfer into a fallback buffer."""

138

139 ref: "Any"

140 keys: "list[FetchKey]"

@@ -562,8 +561,7 @@ class ShardedRDTWeightTransferEngine(

561

562 def _build_static_plan(self, init_info: ShardedRDTWeightTransferInitInfo) -> None:

563 """Build the chunk/free plan once. It never changes across syncs, so

565 - ``update_weights`` needs no per-sync names.

566 - """

564 + ``update_weights`` needs no per-sync names."""

565 if not init_info.group_lens:

566 raise ValueError(

567 "Sharded RDT engine requires init_info.group_lens (the gather-"

@@ -587,8 +585,7 @@ class ShardedRDTWeightTransferEngine(

585 Both sides are sized from the static plan: receive buffers are

586 ``ring_depth`` slots at the largest chunk's ``pack_bytes``, and each bound

587 producer is asked to pre-register a serve ring at the max bytes this

590 - consumer will pull from it. A no-op when this worker has no chunks.

591 - """

588 + consumer will pull from it. A no-op when this worker has no chunks."""

589 plan = self._cached_plan

590 if plan is None or not plan.chunks:

591 return

@@ -633,8 +630,7 @@ class ShardedRDTWeightTransferEngine(

630 ``data_parallel_index``, not ``data_parallel_rank``: vLLM resets the

631 latter to 0 in a dense worker but keeps the former as the distinct global

632 DP rank. Same formula as the sibling ``nccl_engine``, so dense-via-TP and

636 - MoE-via-DP+EP both yield distinct 0..C-1.

637 - """

633 + MoE-via-DP+EP both yield distinct 0..C-1."""

634 pc = self.parallel_config

635 return pc.data_parallel_index * pc.world_size + pc.rank # world_size = TP*PP

636

@@ -642,16 +638,14 @@ class ShardedRDTWeightTransferEngine(

638 """Total inference-worker count. Prefers the driver-supplied

639 ``init_info.num_consumers`` (authoritative -- the driver knows the whole

640 fleet); else ``world_size_across_dp``, the same stride

645 - ``_global_worker_index`` indexes with, so the two agree at any pp.

646 - """

641 + ``_global_worker_index`` indexes with, so the two agree at any pp."""

642 if self._num_consumers_override > 0:

643 return self._num_consumers_override

644 return self.parallel_config.world_size_across_dp

645

646 def start_weight_update(self) -> None:

647 """Put the model's params on meta so layerwise reload streams them in

653 - as each layer's slices land. Baked replay uses checkpoint format.

654 - """

648 + as each layer's slices land. Baked replay uses checkpoint format."""

649 from vllm.model_executor.model_loader.reload import (

650 initialize_layerwise_reload,

651 )

@@ -660,8 +654,7 @@ class ShardedRDTWeightTransferEngine(

654

655 def finish_weight_update(self) -> None:

656 """Drain the deferred pull/process pipeline (so every layer is fully

663 - loaded) before finalizing the layerwise reload.

664 - """

657 + loaded) before finalizing the layerwise reload."""

658 from vllm.model_executor.model_loader.reload import (

659 finalize_layerwise_reload,

660 )

@@ -674,8 +667,7 @@ class ShardedRDTWeightTransferEngine(

667 device sync: post-processing is deferred to background threads and a

668 sync here would block on them and serialize the pull/process pipeline.

669 Completion is guaranteed by ``drain_pending`` in

677 - ``finish_weight_update``.

678 - """

670 + ``finish_weight_update``."""

671 self.receive_weights(self.parse_update_info(update_info))

672

673 def receive_weights(

@@ -717,8 +709,7 @@ class ShardedRDTWeightTransferEngine(

709 device: torch.device,

710 ) -> list[tuple[str, torch.Tensor]]:

711 """Zero-storage lazies for ``names``, dtype/shape from the init metadata,

720 - all feeding the bake's recording sink.

721 - """

712 + all feeding the bake's recording sink."""

713 return [

714 (

715 name,

@@ -986,8 +977,7 @@ class ShardedRDTWeightTransferEngine(

977

978 def _ensure_proc_worker(self) -> None:

979 """Lazily create the per-slot events, the background CUDA stream, the work

989 - queue, and the single processing thread. Idempotent.

990 - """

980 + queue, and the single processing thread. Idempotent."""

981 if self._proc_thread is not None:

982 return

983 import queue

@@ -1046,8 +1036,7 @@ class ShardedRDTWeightTransferEngine(

1036 def _mark_slot_done(self, slot: int) -> None:

1037 """Publish that a queued item's read-done event has been recorded (or the

1038 item failed) so a pull waiting to reuse ``slot`` can proceed to its

1049 - CUDA-event synchronize.

1050 - """

1039 + CUDA-event synchronize."""

1040 with self._slot_cv:

1041 self._slot_done[slot] += 1

1042 self._slot_cv.notify_all()

@@ -1063,8 +1052,7 @@ class ShardedRDTWeightTransferEngine(

1052 """Block until the background thread has processed every queued item and

1053 its stream work is complete, then re-raise any error it hit. Called from

1054 the worker's ``finish_weight_update`` before ``finalize_layerwise_reload``

1066 - so every baked layer is fully loaded (and ``info.reset()``-ed) first.

1067 - """

1055 + so every baked layer is fully loaded (and ``info.reset()``-ed) first."""

1056 if self._proc_queue is not None:

1057 self._proc_queue.join() # every put() item task_done()'d

1058 # The scatter thread feeds the quant thread, so join it SECOND (all

@@ -1116,8 +1104,7 @@ class ShardedRDTWeightTransferEngine(

1104 which is exactly what an owner class is. The cut is a pure function of

1105 the bake and the ownership table. Copy order within a chunk is bake

1106 order, and a module's copies may span chunks (materialize/quant fire on

1119 - its first/last chunk; see ``_build_call_plan``).

1120 - """

1107 + its first/last chunk; see ``_build_call_plan``)."""

1108 assert self._router is not None

1109 by_class: dict[int, list[_Scatter]] = {}

1110 for copies in modules:

@@ -1248,8 +1235,7 @@ class ShardedRDTWeightTransferEngine(

1235 pulls nothing from. Refs are held and drained in ``drain_pending`` so

1236 every signal has EXECUTED before the sync ends: ``begin_sync`` resets

1237 the counters, and a straggler landing in the next sync would credit a

1251 - group it does not belong to.

1252 - """

1238 + group it does not belong to."""

1239 assert self._router is not None

1240 self._pending_frees.extend(self._router.free_group(group_idx))

1241

@@ -1384,8 +1370,7 @@ class ShardedRDTWeightTransferEngine(

1370 the modules' scatters via ``ready``; touches only the scattered params

1371 (never a receive slot), so it can overlap subsequent chunks' RDMA and

1372 scatters. ``info.reset()`` is what makes finalize skip the layer —

1387 - drain_pending joins the quant queue before finalize runs.

1388 - """

1373 + drain_pending joins the quant queue before finalize runs."""

1374 from vllm.model_executor.layers.quantization.base_config import (

1375 QuantizeMethodBase,

1376 )

@@ -1418,8 +1403,7 @@ class ShardedRDTWeightTransferEngine(

1403

1404 def _quant_worker_loop(self) -> None:

1405 """Dedicated quant thread: drains (completed_modules, scatter-done event)

1421 - batches. Errors surface via _proc_error like the scatter thread's.

1422 - """

1406 + batches. Errors surface via _proc_error like the scatter thread's."""

1407 torch.cuda.set_device(self.device)

1408 q = self._quant_queue

1409 assert q is not None

vllm/distributed/weight_transfer/sharded_rdt_fake.py

+2 -4

@@ -206,8 +206,7 @@ class FakeRDTTensor(torch.Tensor):

206

207 def _meta(self) -> torch.Tensor:

208 """A zero-storage meta tensor of this fake's shape/dtype, so PyTorch

209 - itself computes post-op geometry rather than us reimplementing it.

210 - """

209 + itself computes post-op geometry rather than us reimplementing it."""

210 return torch.empty(self.shape, dtype=self.dtype, device="meta")

211

212 @classmethod

@@ -251,8 +250,7 @@ class FakeRDTTensor(torch.Tensor):

250 ):

251 """Append the op and return a child, or a tuple of children for

252 multi-return ops. Each child's geometry comes from running the op on a

254 - meta tensor.

255 - """

253 + meta tensor."""

254 meta = self_._meta()

255 with torch._C.DisableTorchFunctionSubclass():

256 meta_result = func(meta, *args, **kwargs)

vllm/distributed/weight_transfer/sharded_rdt_trainer.py

+14 -28

@@ -132,8 +132,7 @@ PRODUCE_METHOD_NAME = "rdt_produce_weights_batched"

132 @dataclass

133 class _SharedPack:

134 """One generation of one slot-sharing group: the rendezvous state for a

135 - single chunk, identified by the consumers' issue index.

136 - """

135 + single chunk, identified by the consumers' issue index."""

136

137 group: int

138 slot: int

@@ -467,8 +466,7 @@ class _RDTProducerServer:

466

467 def end_sync(self) -> list[int]:

468 """Block until every published group has been freed by its consumers;

470 - return the remaining freed keys so the engine drops its last refs.

471 - """

469 + return the remaining freed keys so the engine drops its last refs."""

470 with self._cache_cond:

471 self._wait_for(

472 lambda: bool(self._inflight_groups),

@@ -481,8 +479,7 @@ class _RDTProducerServer:

479

480 def set_gather_error(self, message: str) -> None:

481 """Record a trainer-side gather failure so blocked serves / publishes

484 - stop waiting and surface it.

485 - """

482 + stop waiting and surface it."""

483 with self._cache_cond:

484 self._gather_error = RuntimeError(message)

485 self._cache_cond.notify_all()

@@ -511,8 +508,7 @@ class _RDTProducerServer:

508 def _new_serve_buffer(self, nbytes: int) -> torch.Tensor:

509 """Allocate + NIXL-register one serve slot: the single allocation seam,

510 so registration cannot be skipped on either of the two paths that make

514 - buffers (the init-time reservation and the serve-path backstop).

515 - """

511 + buffers (the init-time reservation and the serve-path backstop)."""

512 t = torch.empty(nbytes, dtype=torch.uint8, device=self._serve_device)

513 with self._reg_lock:

514 register_nixl_memory(t)

@@ -584,15 +580,13 @@ class _RDTProducerServer:

580 def _share_group(self, consumer_id: int) -> int:

581 """The slot-sharing group ``consumer_id`` belongs to: its index within

582 its own deployment, since that is what fixes the plan. Without a width

587 - the group is the consumer itself and nothing is shared.

588 - """

583 + the group is the consumer itself and nothing is shared."""

584 return consumer_id % self._share_width if self._share_width > 0 else consumer_id

585

586 def _sharers_of(self, sg: int) -> frozenset:

587 """The LIVE consumers of group ``sg``: the rendezvous width. Derived from

588 ``begin_sync``'s live set, so a degraded sync narrows the barrier instead

594 - of waiting forever on a dead deployment. Caller holds ``_cache_cond``.

595 - """

589 + of waiting forever on a dead deployment. Caller holds ``_cache_cond``."""

590 cached = self._sharers.get(sg)

591 if cached is None:

592 if not self._sharing_active or self._live_ids is None:

@@ -628,8 +622,7 @@ class _RDTProducerServer:

622 def _await_shared_pack(self, gen: _SharedPack) -> torch.Tensor:

623 """Block until this generation's packer published its blob, and return

624 it. A failed pack is re-raised here, so every sharer of a bad pack fails

631 - instead of reading a half-written slot.

632 - """

625 + instead of reading a half-written slot."""

626 with self._cache_cond:

627 self._wait_for(

628 lambda: not gen.done and gen.error is None,

@@ -650,8 +643,7 @@ class _RDTProducerServer:

643 """Group ``sg``'s ring slot ``idx``, grown if this chunk outgrew the

644 reservation. Growing registers memory while the fabric is busy, the

645 hazard ``reserve_serve_buffer`` exists to avoid, so it is a backstop: the

653 - reservation is sized from the same static plan as this pack.

654 - """

646 + reservation is sized from the same static plan as this pack."""

647 with self._serve_lock:

648 buffer = self._serve_rings.setdefault(sg, [None] * self._nring)[idx]

649 if buffer is not None and buffer.numel() >= need:

@@ -663,8 +655,7 @@ class _RDTProducerServer:

655

656 def _fail_shared_pack(self, gen: _SharedPack, exc: BaseException) -> None:

657 """Publish a pack failure so waiting sharers raise instead of hanging on

666 - a generation that will never complete.

667 - """

658 + a generation that will never complete."""

659 with self._cache_cond:

660 gen.error = exc

661 self._cache_cond.notify_all()

@@ -866,8 +857,7 @@ class ShardedRDTTrainerWeightTransferEngine(

857 def _rpc(self, method: str, *args: Any) -> Any:

858 """Call one of the server actor's methods and block for the result.

859 The single seam through which the engine talks to its server, so tests

869 - can inject a local (non-Ray) fake server.

870 - """

860 + can inject a local (non-Ray) fake server."""

861 import ray

862

863 return ray.get(getattr(self._server, method).remote(*args))

@@ -944,8 +934,7 @@ class ShardedRDTTrainerWeightTransferEngine(

934 """Fire publish_group WITHOUT waiting on the RPC (the gather loop

935 overlaps the publish's server-side rebuild with the next group's

936 gather) and return a handle that ``_await_publish`` resolves. Ray actor

947 - handle in production; a plain (non-Ray) fake server runs inline.

948 - """

937 + handle in production; a plain (non-Ray) fake server runs inline."""

938 method = self._server.publish_group

939 remote = getattr(method, "remote", None)

940 if remote is not None:

@@ -956,8 +945,7 @@ class ShardedRDTTrainerWeightTransferEngine(

945 """Resolve one async publish so a server-side rebuild error surfaces at

946 window depth instead of at end_sync. Publishes carry nothing back —

947 freed groups flow only through wait_freed/end_sync (one channel; see

959 - publish_group).

960 - """

948 + publish_group)."""

949 import ray

950

951 if isinstance(ref, ray.ObjectRef):

@@ -1144,8 +1132,7 @@ class ShardedRDTTrainerWeightTransferEngine(

1132 by stall watchdog 300s later: consumers route pulls here, the pull

1133 passes the served-names guard, and the cache wait never completes. With

1134 it, that is an immediate error naming the weight. One set lookup per

1147 - name per sync.

1148 - """

1135 + name per sync."""

1136 if self._held_names is None:

1137 return

1138 for name, tensor in zip(names, tensors):

@@ -1346,8 +1333,7 @@ class ShardedRDTTrainerWeightTransferEngine(

1333 credit releases when every live consumer has signaled the group). So

1334 the loop self-paces to the consumers' pull rate with at most

1335 `gather_lookahead + 1` groups resident. Runs on every rank; only the

1349 - sender has an `update_future` to fail fast on.

1350 - """

1336 + sender has an `update_future` to fail fast on."""

1337 assert self.source is not None # guaranteed by trainer_init

1338 # The live IDS ride along with the count: the free barrier needs only

1339 # the count, but the slot-sharing rendezvous needs to know WHICH

vllm/distributed/weight_transfer/sparse_nccl_engine.py

+2 -4

@@ -68,8 +68,7 @@ class SparseNCCLTrainerInitInfo(TrainerInitInfo):

68

69 Same rendezvous shape as the dense NCCL backend (the sender opens its

70 endpoint as NCCL rank 0), but with no packed wire params: sparse transfers

71 - are never packed. `backend` is the factory dispatch key.

72 - """

71 + are never packed. `backend` is the factory dispatch key."""

72

73 backend: ClassVar[str] = "sparse_nccl"

74

@@ -368,8 +367,7 @@ class SparseNCCLTrainerWeightTransferEngine(

367 rebuild or free the patch tensors as soon as a send method returns rather

368 than relying on same-stream ordering. See

369 `NCCLTrainerWeightTransferEngine._post_send_sync` for why there is no

371 - cross-rank barrier.

372 - """

370 + cross-rank barrier."""

371 if torch.cuda.is_available():

372 torch.cuda.current_stream().synchronize()

373

vllm/entrypoints/cli/main.py

+1 -2

@@ -3,8 +3,7 @@

3 """The CLI entrypoints of vLLM.

4

5 Note that all future modules must be lazily loaded within main

6 -to avoid certain eager import breakage.

7 -"""

6 +to avoid certain eager import breakage."""

7

8 import importlib.metadata

9 import sys

vllm/entrypoints/generate/base/serving.py

+1 -2

@@ -295,8 +295,7 @@ class GenerateBaseServing(BaseServing, BeamSearchOnlineMixin):

295 ) -> _T:

296 """Wrap a `create_*` coroutine so that, if it raises or returns an

297 ErrorResponse (i.e. the request never reached the engine), the KV

298 - connector is notified to free any pinned remote-prefill blocks.

299 - """

298 + connector is notified to free any pinned remote-prefill blocks."""

299 kv_transfer_params = self.has_kv_connector and request.kv_transfer_params

300 if not kv_transfer_params or not kv_transfer_params.get("do_remote_prefill"):

301 return await awaitable

vllm/entrypoints/openai/models/serving.py

+2 -4

@@ -123,8 +123,7 @@ class OpenAIServingModels:

123

124 async def init_static_loras(self):

125 """Loads all static LoRA modules.

126 - Raises if any fail to load

127 - """

126 + Raises if any fail to load"""

127 if self.static_lora_modules is None:

128 return

129 for lora in self.static_lora_modules:

@@ -149,8 +148,7 @@ class OpenAIServingModels:

148

149 async def show_available_models(self) -> ModelList:

150 """Show available models. This includes the base model and all

152 - adapters.

153 - """

151 + adapters."""

152 model_list = await self.registry.show_available_models()

153 lora_cards = [

154 ModelCard(

vllm/entrypoints/openai/responses/context.py

+1 -2

@@ -408,8 +408,7 @@ class ParsableContext(ConversationContext):

408

409 def need_builtin_tool_call(self) -> bool:

410 """Return true if the last message is a builtin tool call

411 - that the request has enabled.

412 - """

411 + that the request has enabled."""

412 last_message = self.response_messages[-1]

413 if last_message.type != "function_call":

414 return False

vllm/entrypoints/openai/responses/harmony.py

+1 -2

@@ -52,8 +52,7 @@ logger = init_logger(__name__)

52

53 def _parse_harmony_format_message(chat_msg: dict) -> Message:

54 """Reconstruct a Message from Harmony-format dict,

55 - preserving channel, recipient, and content_type.

56 - """

55 + preserving channel, recipient, and content_type."""

56 author_dict = chat_msg["author"]

57 role = author_dict.get("role")

58 name = author_dict.get("name")

vllm/entrypoints/openai/responses/protocol.py

+1 -2

@@ -117,8 +117,7 @@ def serialize_messages(msgs):

117

118 class ResponseRawMessageAndToken(OpenAIBaseModel):

119 """Class to show the raw message.

120 - If message / tokens diverge, tokens is the source of truth

121 - """

120 + If message / tokens diverge, tokens is the source of truth"""

121

122 message: str

123 tokens: list[int]

vllm/entrypoints/pooling/pooling/io_processor.py

+1 -2

@@ -44,8 +44,7 @@ class PluginWithoutIOProcessorPlugins(PoolingIOProcessor):

44

45 class PluginWithIOProcessorPlugins(PoolingIOProcessor):

46 """IO Processor plugins are a feature that allows pre- and post-processing

47 - of the model input and output for pooling models.

48 - """

47 + of the model input and output for pooling models."""

48

49 name = "plugin"

50

vllm/entrypoints/scale_out/token_in_token_out/mm_serde.py

+1 -2

@@ -1,8 +1,7 @@

1 # SPDX-License-Identifier: Apache-2.0

2 # SPDX-FileCopyrightText: Copyright contributors to the vLLM project

3 """Encode/decode utilities for multimodal tensors and field metadata

4 -over JSON/HTTP, used by the disaggregated generate endpoint.

5 -"""

4 +over JSON/HTTP, used by the disaggregated generate endpoint."""

5

6 from __future__ import annotations

7

vllm/entrypoints/serve/exception_handling/handlers/validation.py

+1 -2

@@ -126,8 +126,7 @@ def _format_error(err: object) -> str:

126

127 def _is_internal_loc_segment(segment: str) -> bool:

128 """True if `segment` is a Pydantic-internal wrapper/union-branch

129 - marker rather than a user-meaningful field name or list index.

130 - """

129 + marker rather than a user-meaningful field name or list index."""

130 if _BRACKETED_INTERNAL_RE.search(segment):

131 return True

132 return segment.lower() in _INTERNAL_LOC_MARKERS

vllm/entrypoints/serve/lora/api_router.py

+2 -4

@@ -63,8 +63,7 @@ def _attach_router(app: FastAPI):

63 @router.post("/v1/load_lora_adapter", dependencies=[Depends(validate_json_request)])

64 async def load_lora_adapter(request: LoadLoRAAdapterRequest, raw_request: Request):

65 """Handle POST /v1/load_lora_adapter: load a LoRA adapter into the

66 - serving engine.

67 - """

66 + serving engine."""

67 handler: OpenAIServingModels = models(raw_request)

68 response = await handler.load_lora_adapter(request)

69 if isinstance(response, ErrorResponse):

@@ -86,8 +85,7 @@ def _attach_router(app: FastAPI):

85 request: UnloadLoRAAdapterRequest, raw_request: Request

86 ):

87 """Handle POST /v1/unload_lora_adapter: unload a LoRA adapter from

89 - the serving engine.

90 - """

88 + the serving engine."""

89 handler: OpenAIServingModels = models(raw_request)

90 response = await handler.unload_lora_adapter(request)

91 if isinstance(response, ErrorResponse):

vllm/entrypoints/serve/sagemaker/api_router.py

+2 -4

@@ -30,8 +30,7 @@ EndpointFn = Callable[[RequestType, Request], Awaitable[Any]]

30 def _snapshot_handler_levels() -> list[tuple[logging.Handler, int]]:

31 """Snapshot handler levels for loggers that may be affected by third-party

32 logging configuration side effects (e.g. model_hosting_container_standards

33 - calling configure_root_logger() at import time).

34 - """

33 + calling configure_root_logger() at import time)."""

34 loggers = [logging.getLogger(), logging.getLogger("vllm")]

35 seen: set[int] = set()

36 snapshot: list[tuple[logging.Handler, int]] = []

@@ -74,8 +73,7 @@ def _attach_router(

73 model_config: ModelConfig | None = None,

74 ):

75 """Register the SageMaker hosting routes (/ping, /invocations) on the

77 - app.

78 - """

76 + app."""

77 import model_hosting_container_standards.sagemaker as sagemaker_standards

78

79 router = APIRouter()

vllm/entrypoints/serve/utils/fingerprint.py

+2 -4

@@ -33,8 +33,7 @@ def set_default_fingerprint_mode(

33 custom_value: str | None = None,

34 ) -> None:

35 """Configure the fingerprint mode for subsequent ``get_system_fingerprint``

36 - calls. Called once at server startup.

37 - """

36 + calls. Called once at server startup."""

37 global _DEFAULT_MODE, _CUSTOM_VALUE

38 _DEFAULT_MODE = mode

39 _CUSTOM_VALUE = custom_value

@@ -42,8 +41,7 @@ def set_default_fingerprint_mode(

41

42 def get_system_fingerprint(vllm_config: Any) -> str | None:

43 """Return the fingerprint for ``vllm_config`` using the mode configured by

45 - ``set_default_fingerprint_mode``.

46 - """

44 + ``set_default_fingerprint_mode``."""

45 return build_system_fingerprint(vllm_config, _DEFAULT_MODE, _CUSTOM_VALUE)

46

47

vllm/entrypoints/speech_to_text/base/serving.py

+2 -4

@@ -92,8 +92,7 @@ def asr_inter_chunk_separator(

92

93 class SpeechToTextBaseServing(GenerateBaseServing):

94 """Base class for speech-to-text operations like transcription and

95 - translation.

96 - """

95 + translation."""

96

97 def __init__(

98 self,

@@ -427,8 +426,7 @@ class SpeechToTextBaseServing(GenerateBaseServing):

426 stream_generator_method: Callable[..., AsyncGenerator[str, None]],

427 ) -> T | V | AsyncGenerator[str, None] | ErrorResponse:

428 """Base method for speech-to-text operations like transcription and

430 - translation.

431 - """

429 + translation."""

430 if request.stream and request.use_beam_search:

431 return self.create_error_response(

432 "Streaming is not currently supported with beam search"

vllm/env_override.py

+1 -2

@@ -581,8 +581,7 @@ _constrain_to_fx_strides_patched = False

581 def _apply_constrain_to_fx_strides_patch():

582 """Patch lowering.constrain_to_fx_strides globally. Safe to call

583 multiple times; only the first call does anything.

584 - Only applies for torch >= 2.11 and < 2.12.

585 - """

584 + Only applies for torch >= 2.11 and < 2.12."""

585 global _constrain_to_fx_strides_patched

586 if _constrain_to_fx_strides_patched:

587 return

vllm/envs.py

+1 -2

@@ -2249,8 +2249,7 @@ def compile_factors() -> dict[str, object]:

2249 """Return env vars used for torch.compile cache keys.

2250

2251 Start with every known vLLM env var; drop entries in `ignored_factors`;

2252 - hash everything else. This keeps the cache key aligned across workers.

2253 - """

2252 + hash everything else. This keeps the cache key aligned across workers."""

2253 ignored_factors: set[str] = {

2254 "MAX_JOBS",

2255 "VLLM_RPC_BASE_PATH",

vllm/forward_context.py

+1 -2

@@ -209,8 +209,7 @@ def is_forward_context_available() -> bool:

209

210 def in_piecewise_cudagraph() -> bool:

211 """Whether the current forward runs in piecewise cudagraph mode (graph

212 - segments separated by eager breaks), at capture or replay time.

213 - """

212 + segments separated by eager breaks), at capture or replay time."""

213 return (

214 is_forward_context_available()

215 and get_forward_context().cudagraph_runtime_mode == CUDAGraphMode.PIECEWISE

vllm/logger.py

+1 -2

@@ -216,8 +216,7 @@ def _configure_vllm_root_logger() -> None:

216 def init_logger(name: str) -> _VllmLogger:

217 """The main purpose of this function is to ensure that loggers are

218 retrieved in such a way that we can be sure the root vllm logger has

219 - already been configured.

220 - """

219 + already been configured."""

220 logger = logging.getLogger(name)

221

222 for method_name, method in _METHODS_TO_PATCH.items():

vllm/lora/layers/column_parallel_linear.py

+1 -2

@@ -733,8 +733,7 @@ class MergedColumnParallelLinearVariableSliceWithLoRA(

733 lora_b: torch.Tensor | list[torch.Tensor],

734 ):

735 """Override to handle single tensor weights

736 - that need to be split into slices.

737 - """

736 + that need to be split into slices."""

737 self.reset_lora(index)

738

739 # Handle case where checkpoint has single tensor weights

vllm/lora/lora_model.py

+1 -2

@@ -90,8 +90,7 @@ class LoRAModel:

90 def clone(self, lora_model_id: int) -> "LoRAModel":

91 """Return a copy of the object with different ids.

92

93 - Will share the underlying tensors.

94 - """

93 + Will share the underlying tensors."""

94 return self.__class__(

95 lora_model_id,

96 rank=self.rank,

vllm/lora/worker_manager.py

+3 -6

@@ -28,8 +28,7 @@ class WorkerLoRAManager:

28 """WorkerLoRAManager that manages LoRA models on the worker side.

29

30 Every request, the requested LoRAs will be loaded (unless they are already

31 - loaded), and every other LoRA will be unloaded.

32 - """

31 + loaded), and every other LoRA will be unloaded."""

32

33 _manager_cls: type[LoRAModelManager] = LoRAModelManager

34

@@ -75,8 +74,7 @@ class WorkerLoRAManager:

74 @contextmanager

75 def dummy_lora_cache(self):

76 """Use this context manager to reuse the dummy lora model

78 - to avoid creating it repeatedly.

79 - """

77 + to avoid creating it repeatedly."""

78 self._cached_dummy_lora = None

79 yield

80 self._cached_dummy_lora = False

@@ -251,8 +249,7 @@ class LRUCacheWorkerLoRAManager(WorkerLoRAManager):

249

250 Uses an LRU Cache. Every request, the requested LoRAs will be loaded

251 (unless they are already loaded) and least recently used LoRAs will

254 - be unloaded if the cache is above capacity.

255 - """

252 + be unloaded if the cache is above capacity."""

253

254 _manager_cls: type[LRUCacheLoRAModelManager] = LRUCacheLoRAModelManager

255

vllm/model_executor/hw_agnostic/custom_op.py

+1 -2

@@ -135,8 +135,7 @@ class CustomOp(nn.Module):

135

136 def forward_native(self, *args, **kwargs):

137 """PyTorch-native implementation; OOT plugins override via

138 - :meth:`forward_oot`.

139 - """

138 + :meth:`forward_oot`."""

139 raise NotImplementedError

140

141 def forward_oot(self, *args, **kwargs):

vllm/model_executor/hw_agnostic/layers/layernorm.py

+1 -2

@@ -10,8 +10,7 @@ from vllm.model_executor.hw_agnostic.custom_op import CustomOp

10 @CustomOp.register("rms_norm")

11 class RMSNorm(CustomOp):

12 """``x -> w * x / sqrt(E[x^2] + eps)``. With ``residual``, fuses

13 - ``residual += x`` then RMSNorm and returns ``(normalized, residual)``.

14 - """

13 + ``residual += x`` then RMSNorm and returns ``(normalized, residual)``."""

14

15 def __init__(

16 self,

vllm/model_executor/kernels/linear/__init__.py

+4 -8

@@ -881,8 +881,7 @@ def choose_mp_linear_kernel(

881

882 def init_mxfp8_linear_kernel(*, bmm_batch_size: int | None = None) -> Mxfp8LinearKernel:

883 """Select and instantiate the best MXFP8 linear kernel for the

884 - current platform.

885 - """

884 + current platform."""

885 config = Mxfp8LinearLayerConfig(bmm_batch_size=bmm_batch_size)

886

887 platform = current_platform._enum

@@ -934,8 +933,7 @@ def init_mxfp4_linear_kernel(

933 activation_quant_key: QuantKey | None = None,

934 ) -> MxFp4LinearKernel:

935 """Select and instantiate the best MXFP4 linear kernel for the

937 - current platform.

938 - """

936 + current platform."""

937 config = MxFp4LinearLayerConfig(

938 activation_quant_key=activation_quant_key,

939 )

@@ -982,8 +980,7 @@ def init_mxfp6_linear_kernel(

980 activation_quant_key: QuantKey | None = None,

981 ) -> MxFp6LinearKernel:

982 """Select and instantiate the best MXFP6 linear kernel for the

985 - current platform.

986 - """

983 + current platform."""

984 config = MxFp6LinearLayerConfig(

985 weight_quant_key=weight_quant_key,

986 activation_quant_key=activation_quant_key,

@@ -1071,8 +1068,7 @@ def init_wfp8_a16_linear_kernel(

1068

1069 def init_nvfp4_linear_kernel(use_a16: bool = False) -> NvFp4LinearKernel:

1070 """Select and instantiate the best NVFP4 linear kernel for the

1074 - current platform.

1075 - """

1071 + current platform."""

1072 config = NvFp4LinearLayerConfig()

1073 a16_kernels = (

1074 FlashInferCuteDslNvFp4W4A16LinearKernel,

vllm/model_executor/kernels/linear/nvfp4/flashinfer.py

+1 -2

@@ -177,8 +177,7 @@ class FlashInferCutlassNvFp4LinearKernel(NvFp4LinearKernel):

177

178 def input_quant_key(self) -> QuantKey | None:

179 """This kernel supports dynamic quantization of the input. By

180 - convention, pre-quantized blockscales must use the swizzled layout.

181 - """

180 + convention, pre-quantized blockscales must use the swizzled layout."""

181 return kNvfp4Dynamic

182

183 @classmethod

vllm/model_executor/kernels/linear/scaled_mm/BlockScaledMMLinearKernel.py

+1 -2

@@ -41,8 +41,7 @@ class FP8BlockParams(FP8Params):

41 @property

42 def block_scale_attr(self) -> str:

43 """Fp8LinearMethod registers the block scale as ``weight_scale_inv``,

44 - compressed-tensors as ``weight_scale``.

45 - """

44 + compressed-tensors as ``weight_scale``."""

45 return (

46 self.WEIGHT_SCALE

47 if self.weight_scale_inv is None

vllm/model_executor/kernels/linear/scaled_mm/cutlass.py

+2 -4

@@ -174,8 +174,7 @@ class CutlassFP8ScaledMMLinearKernel(FP8ScaledMMLinearKernel):

174

175 def input_quant_key(self) -> QuantKey | None:

176 """Only static per-tensor activation quantization is supported for external

177 - quantization.

178 - """

177 + quantization."""

178 if self.config.activation_quant_key == kFp8StaticTensorSym:

179 return kFp8StaticTensorSym

180 return None

@@ -185,8 +184,7 @@ class CutlassFP8ScaledMMLinearKernel(FP8ScaledMMLinearKernel):

184 x: torch.Tensor, dim: int, alignment: int, value: float = 0.0

185 ) -> torch.Tensor:

186 """Pad tensor ``x`` along ``dim`` to the next multiple of

188 - ``alignment``.

189 - """

187 + ``alignment``."""

188 remainder = x.shape[dim] % alignment

189 if remainder == 0:

190 return x

vllm/model_executor/layers/activation.py

+1 -2

@@ -804,8 +804,7 @@ _ACTIVATION_REGISTRY = LazyDict(

804

805 def _get_gelu_pytorch_tanh() -> nn.Module:

806 """Get PyTorch GELU with tanh approximation, with ROCm fallback

807 - and fast GELU for ARM.

808 - """

807 + and fast GELU for ARM."""

808 if current_platform.is_rocm():

809 # TODO:[ROCm] PyTorch native GELU with tanh is unstable with torch.compile

810 logger.warning_once(

vllm/model_executor/layers/attention/mla_attention.py

+1 -2

@@ -1568,8 +1568,7 @@ class MLACommonBackend(AttentionBackend):

1568 @classmethod

1569 def customize_spec(cls, spec: "AttentionSpec") -> "AttentionSpec":

1570 """Per-token-head modes pack an inline fp32 scale pair after the

1571 - latent data (single-sided: ``head_size_v == 0`` for MLA).

1572 - """

1571 + latent data (single-sided: ``head_size_v == 0`` for MLA)."""

1572 spec = super().customize_spec(spec)

1573 mode = spec.kv_quant_mode

1574 if spec.state_content_bytes is not None or not mode.is_per_token_head:

vllm/model_executor/layers/fused_allreduce_gemma_rms_norm.py

+1 -2

@@ -55,8 +55,7 @@ _FI_SUPPORTED_DTYPES = (torch.bfloat16, torch.float16)

55

56 def _max_token_num(tp_size: int, hidden_size: int, dtype: torch.dtype) -> int | None:

57 """Workspace token budget for flashinfer fused all-reduce, or None if the

58 - current world size / device is unsupported. Mirrors ``FlashInferAllReduce``.

59 - """

58 + current world size / device is unsupported. Mirrors ``FlashInferAllReduce``."""

59 from vllm.config.compilation import PassConfig

60

61 max_size_mb = PassConfig.default_fi_allreduce_fusion_max_size_mb().get(tp_size)

vllm/model_executor/layers/fused_embed_norm.py

+1 -2

@@ -170,8 +170,7 @@ def _fused_embed_eh_norm_kernel(

170 ``table[ids]``, zero it at position 0, RMSNorm(embed) with enorm and

171 RMSNorm(prev_hidden) with hnorm, written side-by-side into ``out`` ([N, 2H])

172 ready for the eh_proj GEMM. Replaces embedding lookup + where + 2x RMSNorm +

173 - cat. Requires the full table on-rank (replicated embedding).

174 - """

173 + cat. Requires the full table on-rank (replicated embedding)."""

174 tok = tl.program_id(0)

175 off = tl.arange(0, BLOCK)

176 mask = off < H

vllm/model_executor/layers/fused_moe/activation.py

+1 -2

@@ -56,8 +56,7 @@ class MoEActivation(Enum):

56 @property

57 def custom_op_name(self) -> str:

58 """Maps to the CustomOp name of activations

59 - in vllm/model_executor/layers/activation.py.

60 - """

59 + in vllm/model_executor/layers/activation.py."""

60 return _CUSTOM_OP_NAMES[self]

61

62 def without_mul(self) -> "MoEActivation":

vllm/model_executor/layers/fused_moe/experts/aiter_mxfp8_moe.py

+1 -2

@@ -36,8 +36,7 @@ def is_aiter_mxfp8_moe_available() -> bool:

36 nightly lacking #3811 would wrongly select FlyDSL instead of falling back to

37 the native Triton dot_scaled path. #3811 added no probe-able public symbol,

38 so detect the ``minimax_m3_mxfp8`` tuned config it shipped. Every check fails

39 - closed (returns False -> triton dot_scaled), which is always safe.

40 - """

39 + closed (returns False -> triton dot_scaled), which is always safe."""

40 if not (current_platform.is_rocm() and current_platform.supports_mx()):

41 return False

42 try:

vllm/model_executor/layers/fused_moe/experts/cpu_moe.py

+3 -6

@@ -192,8 +192,7 @@ class CPUUnquantizedExperts(mk.FusedMoEExpertsModular):

192

193 def _pad_moe_intermediate(self, layer: torch.nn.Module) -> None:

194 """Zero-pad the per-partition MoE intermediate dim of both weights and

195 - the expert bias, see `_padded_intermediate_size`.

196 - """

195 + the expert bias, see `_padded_intermediate_size`."""

196 intermediate_size = self.moe_config.intermediate_size_per_partition

197 padded_size = self._padded_intermediate_size(self.moe_config)

198 if padded_size == intermediate_size:

@@ -953,8 +952,7 @@ class CPUExpertsInt4(mk.FusedMoEExpertsModular):

952

953 class CPUExpertsInt8(mk.FusedMoEExpertsModular):

954 """CPU INT8 W8A8 per-channel weight / dynamic per-token activation

956 - modular MoE experts.

957 - """

955 + modular MoE experts."""

956

957 def __init__(

958 self,

@@ -1287,8 +1285,7 @@ class ArmCPUExpertsInt8(mk.FusedMoEExpertsModular):

1285

1286 class ZenCPUExpertsInt8(mk.FusedMoEExpertsModular):

1287 """AMD Zen INT8 MoE with per-token activation and channelwise weight

1290 - quantization, dispatched through zentorch.

1291 - """

1288 + quantization, dispatched through zentorch."""

1289

1290 def __init__(

1291 self,

vllm/model_executor/layers/fused_moe/experts/mxfp8_native_moe.py

+1 -2

@@ -34,8 +34,7 @@ from vllm.triton_utils import tl, triton

34

35 def _select_cfg(M, N, K, block_m):

36 """Pick the launch config from host constants only (M=num_valid_tokens, N, K,

37 - block_m) — graph-capture safe (no GPU-scalar branch).

38 - """

37 + block_m) — graph-capture safe (no GPU-scalar branch)."""

38 # Per-regime winners (measured, isolated cuda-event A/B on gfx950, GPU 3):

39 # BLOCK_K=256 (fewer K-iters + bigger MX scale-load coalesced with the dot),

40 # num_stages=2 (software-pipeline overlaps the E8M0 scale-load with the scaled

vllm/model_executor/layers/fused_moe/oracle/base.py

+5 -10

@@ -55,28 +55,24 @@ class MoEKernelOracle(ABC, Generic[BackendT]):

55 @abstractmethod

56 def backend_enum_cls(self) -> type[BackendT]:

57 """Return the concrete `Enum` class enumerating this oracle's

58 - backends (e.g. `UnquantizedMoeBackend`, `Fp8MoeBackend`).

59 - """

58 + backends (e.g. `UnquantizedMoeBackend`, `Fp8MoeBackend`)."""

59

60 @abstractmethod

61 def get_priority_backends(self, moe_config: FusedMoEConfig) -> list[BackendT]:

62 """Return platform-appropriate backends in priority order for

64 - this `moe_config`.

65 - """

63 + this `moe_config`."""

64

65 @abstractmethod

66 def backend_to_kernel_cls(

67 self, backend: BackendT

68 ) -> list[type[mk.FusedMoEExperts]]:

69 """Map a backend enum value to its concrete `FusedMoEExperts`

72 - subclasses, in selection priority order.

73 - """

70 + subclasses, in selection priority order."""

71

72 @abstractmethod

73 def map_backend(self, runner_backend: MoEBackend) -> BackendT:

74 """Map a user-facing `MoEBackend` (from the runner config) to

78 - this oracle's enum.

79 - """

75 + this oracle's enum."""

76

77 @abstractmethod

78 def select_backend(

@@ -108,8 +104,7 @@ class MoEKernelOracle(ABC, Generic[BackendT]):

104 routing_tables: tuple[torch.Tensor, torch.Tensor, torch.Tensor] | None = None,

105 ) -> mk.FusedMoEKernel:

106 """Construct the `FusedMoEKernel` (Prepare/Finalize + Experts

111 - combinator) for the chosen backend.

112 - """

107 + combinator) for the chosen backend."""

108

109 def convert_to_kernel_format(

110 self,

vllm/model_executor/layers/fused_moe/oracle/fp8.py

+1 -2

@@ -451,8 +451,7 @@ def _humming_fp8_weight_schema(

451 layer: RoutedExperts, weight: torch.Tensor, weight_scale: torch.Tensor

452 ) -> dict[str, Any]:

453 """Build the humming weight schema from the canonical on-device fp8/mxfp8

454 - tensors (scale dtype/shape, block size), not the producing quant method.

455 - """

454 + tensors (scale dtype/shape, block size), not the producing quant method."""

455 # mxfp8: e8m0 group-32 scales (stored as uint8 bytes or e8m0). humming has

456 # no compressed-tensors mxfp8 loader; its modelopt schema fits both sources.

457 if weight_scale.dtype in (torch.uint8, torch.float8_e8m0fnu):

vllm/model_executor/layers/fused_moe/oracle/int8.py

+1 -2

@@ -224,8 +224,7 @@ def _humming_int8_weight_schema(

224 weight: torch.Tensor, weight_scale: torch.Tensor

225 ) -> dict[str, Any]:

226 """Build the humming compressed-tensors int8 schema from the canonical

227 - on-device tensors; humming does the signed-int8 -> native conversion.

228 - """

227 + on-device tensors; humming does the signed-int8 -> native conversion."""

228 config: dict[str, Any] = {

229 "quant_method": "compressed-tensors",

230 "format": "int-quantized",

vllm/model_executor/layers/fused_moe/oracle/int_wna16.py

+2 -4

@@ -512,8 +512,7 @@ def _process_weights_flashinfer(

512

513 def _pad_w13_shard_cols(x: torch.Tensor, unit: int, padded_unit: int) -> torch.Tensor:

514 """Zero-pad each of the two gate/up shards of a ``(E, rows, 2 * unit)``

515 - tensor along its last dim, from ``unit`` to ``padded_unit`` columns.

516 - """

515 + tensor along its last dim, from ``unit`` to ``padded_unit`` columns."""

516 if padded_unit == unit:

517 return x

518 e, rows, _ = x.shape

@@ -1081,8 +1080,7 @@ def _humming_wna16_weight_schema(

1080 quant_config: QuantizationConfig | QuantizationArgs | None,

1081 ) -> dict[str, Any]:

1082 """Humming weight schema for a WNA16 checkpoint, derived from the quant

1084 - config rather than the running kernel.

1085 - """

1083 + config rather than the running kernel."""

1084 from vllm.model_executor.layers.quantization.auto_awq import AutoAWQConfig

1085 from vllm.model_executor.layers.quantization.auto_gptq import AutoGPTQConfig

1086

vllm/model_executor/layers/fused_moe/oracle/mxfp4.py

+3 -6

@@ -331,8 +331,7 @@ def map_mxfp4_backend(runner_backend: MoEBackend) -> list[Mxfp4MoeBackend]:

331

332 def _get_priority_backends_for_gpt_oss() -> list[Mxfp4MoeBackend]:

333 """Available backends in priority order, BF16-act variant before

334 - activation-quantized variant within each vendor family.

335 - """

334 + activation-quantized variant within each vendor family."""

335 _AVAILABLE_BACKENDS = [

336 Mxfp4MoeBackend.FLASHINFER_TRTLLM_MXFP4_BF16,

337 Mxfp4MoeBackend.FLASHINFER_TRTLLM_MXFP4_MXFP8,

@@ -411,8 +410,7 @@ def _resolve_activation_key(

410 model_activation_key: QuantKey | None,

411 ) -> QuantKey | None:

412 """Combine the model-supplied activation key with the user override.

414 - Raises on conflict (both set and disagreeing).

415 - """

413 + Raises on conflict (both set and disagreeing)."""

414 user_override = _user_moe_activation_override()

415 if user_override is None:

416 return model_activation_key

@@ -462,8 +460,7 @@ def _filter_by_activation(

460 ) -> list[Mxfp4MoeBackend]:

461 """Pick variants matching ``requested_activation_key``; without one,

462 prefer BF16 if the list has any, else keep the list as-is so explicit

465 - non-BF16 picks (e.g. the ``_afp8`` aliases) still land.

466 - """

463 + non-BF16 picks (e.g. the ``_afp8`` aliases) still land."""

464 if requested_activation_key is not None:

465 return [

466 b

vllm/model_executor/layers/fused_moe/router/zero_expert_router.py

+1 -2

@@ -71,8 +71,7 @@ class ZeroExpertRouter(BaseRouter):

71 input_ids: torch.Tensor | None = None,

72 ) -> tuple[torch.Tensor, torch.Tensor]:

73 """Compute routing with full bias, compute zero expert output,

74 - mask zero expert IDs.

75 - """

74 + mask zero expert IDs."""

75 topk_weights, topk_ids = fused_topk_bias(

76 hidden_states=hidden_states,

77 gating_output=router_logits,

vllm/model_executor/layers/indexer_topk.py

+3 -6

@@ -163,8 +163,7 @@ class SparseIndexerTopk(torch.nn.Module):

163 ) -> str:

164 """Resolve the decode top-k implementation from the configured

165 backend ("auto" = the pre-existing chain, or a validated explicit

166 - value).

167 - """

166 + value)."""

167 if self._backend == "auto":

168 return self._resolve_auto(logits, topk_tokens, num_rows)

169

@@ -211,8 +210,7 @@ class SparseIndexerTopk(torch.nn.Module):

210 self, logits: torch.Tensor, topk_tokens: int, num_rows: int

211 ) -> str:

212 """The pre-existing priority chain: cooperative -> persistent ->

214 - per_row. deep_select/flashinfer/torch are opt-in only.

215 - """

213 + per_row. deep_select/flashinfer/torch are opt-in only."""

214 if not self._cooperative_constraints(logits, topk_tokens, num_rows):

215 return "cooperative"

216 if self._is_cuda and topk_tokens in (512, 1024, 2048):

@@ -274,8 +272,7 @@ class SparseIndexerTopk(torch.nn.Module):

272 max_seq_len: int,

273 ) -> None:

274 """Run the resolved decode top-k implementation, writing into

277 - topk_indices (int32, -1 fill for rows shorter than topk_tokens).

278 - """

275 + topk_indices (int32, -1 fill for rows shorter than topk_tokens)."""

276 backend = self.resolve_backend(logits, topk_tokens, logits.shape[0])

277 if backend == "deep_select":

278 row_ends = self._row_ends(seq_lens, next_n, logits.shape[0])

vllm/model_executor/layers/linear.py

+1 -2

@@ -158,8 +158,7 @@ class LinearMethodBase(QuantizeMethodBase):

158 bias: torch.Tensor | None = None,

159 ) -> torch.Tensor:

160 """Apply the weights in layer to the input tensor.

161 - Expects create_weights to have been called before on the layer.

162 - """

161 + Expects create_weights to have been called before on the layer."""

162 raise NotImplementedError

163

164

vllm/model_executor/layers/mamba/gdn/qwen_gdn_linear_attn.py

+1 -2

@@ -861,8 +861,7 @@ class QwenGatedDeltaNetAttention(GatedDeltaNetAttention):

861 hidden_states: torch.Tensor,

862 ) -> torch.Tensor:

863 """ROCm forward using AITER Triton fused projection+attention when

864 - available, otherwise falling back to the generic CUDA path.

865 - """

864 + available, otherwise falling back to the generic CUDA path."""

865 if GDN_AITER_TRITON_AVAILABLE:

866 num_tokens = hidden_states.size(0)

867 projected_states_qkvz, _ = self.in_proj_qkvz(hidden_states)

vllm/model_executor/layers/mamba/mamba_utils.py

+1 -2

@@ -261,8 +261,7 @@ class MambaStateShapeCalculator:

261 @classmethod

262 def extra_groups_for_head_shards(cls, ngroups: int, tp_size: int):

263 """Compute the increase in group numbers to account for

264 - replication in order to accompany the head shards.

265 - """

264 + replication in order to accompany the head shards."""

265 # in the case ngoups % tp_size == 0, this will be zero

266 if ngroups % tp_size == 0:

267 return 0

vllm/model_executor/layers/mamba/ops/mamba_ssm.py

+2 -4

@@ -59,8 +59,7 @@ def get_ssm_device_name() -> str:

59

60 def _canonical_cache_dtype(cache_dtype: str) -> str:

61 """Canonical key for config lookup. bf16 and fp16 share the same tuned

62 - configs because the kernel only sees bit width when accessing state.

63 - """

62 + configs because the kernel only sees bit width when accessing state."""

63 return "float16" if cache_dtype == "bfloat16" else cache_dtype

64

65

@@ -161,8 +160,7 @@ _ssm_config_override: tuple[int, int] | None = None

160 @contextmanager

161 def override_ssm_config(config: tuple[int, int]):

162 """Pin ``try_get_optimal_ssm_config`` to ``config`` for the duration of

164 - the context. Used by the tuning benchmark to time specific configs.

165 - """

163 + the context. Used by the tuning benchmark to time specific configs."""

164 global _ssm_config_override

165 prev = _ssm_config_override

166 _ssm_config_override = config

vllm/model_executor/layers/minimax_rms_norm/lamport_workspace.py

+1 -2

@@ -210,8 +210,7 @@ class LamportWorkspace:

210 @property

211 def workspace(self) -> torch.Tensor:

212 """Device tensor (int64) that can be passed to the kernel

213 - as ``void** workspace``.

214 - """

213 + as ``void** workspace``."""

214 return self._workspace

215

216 # ------------------------------------------------------------------

vllm/model_executor/layers/quantization/base_config.py

+4 -8

@@ -46,24 +46,21 @@ class QuantizeMethodBase(ABC):

46 ):

47 """Create weights for a layer.

48

49 - The weights will be set as attributes of the layer.

50 - """

49 + The weights will be set as attributes of the layer."""

50 raise NotImplementedError

51

52 @abstractmethod

53 def apply(self, layer: torch.nn.Module, *args, **kwargs) -> torch.Tensor:

54 """Apply the weights in layer to the input tensor.

55

57 - Expects create_weights to have been called before on the layer.

58 - """

56 + Expects create_weights to have been called before on the layer."""

57 raise NotImplementedError

58

59 # Not required functions

60 def embedding(self, layer: torch.nn.Module, *args, **kwargs) -> torch.Tensor:

61 """Gather embeddings in the layer based on indices in the input tensor.

62

65 - Expects create_weights to have been called before on the layer.

66 - """

63 + Expects create_weights to have been called before on the layer."""

64 raise NotImplementedError

65

66 # Not required functions

@@ -76,8 +73,7 @@ class QuantizeMethodBase(ABC):

73 Quantization methods that need special weight handling (e.g. repacked

74 weights) override this.

75

79 - Expects create_weights to have been called before on the layer.

80 - """

76 + Expects create_weights to have been called before on the layer."""

77 layer.weight = embed_tokens.weight

78 return layer

79

vllm/model_executor/layers/quantization/compressed_tensors/compressed_tensors.py

+2 -4

@@ -665,8 +665,7 @@ class CompressedTensorsConfig(QuantizationConfig):

665 ) -> bool:

666 """Weight N-bit INT (pack-quantized for sub-byte, int-quantized for 8-bit)

667 with static per-tensor INT8 input/output activation quant, applied as a float

668 - fake-quant around a weight-only matmul.

669 - """

668 + fake-quant around a weight-only matmul."""

669 is_int_pack_format = format in (

670 CompressionFormat.pack_quantized.value,

671 CompressionFormat.int_quantized.value,

@@ -702,8 +701,7 @@ class CompressedTensorsConfig(QuantizationConfig):

701 format: str | None,

702 ) -> bool:

703 """Weight N-bit INT with symmetric dynamic INT activation quant

705 - via Humming kernel.

706 - """

704 + via Humming kernel."""

705 if input_quant is None:

706 return False

707 is_pack_format = format == CompressionFormat.pack_quantized.value

vllm/model_executor/layers/quantization/compressed_tensors/compressed_tensors_embedding.py

+1 -2

@@ -37,8 +37,7 @@ def _dequant_gather_kernel(

37 BLOCK: tl.constexpr,

38 ):

39 """Gather embedding rows by token id, unpack int32-packed INT weights, and

40 - dequantize to ``out`` dtype in one pass (no int8 intermediate).

41 - """

40 + dequantize to ``out`` dtype in one pass (no int8 intermediate)."""

41 row = tl.program_id(0)

42 col = tl.program_id(1) * BLOCK + tl.arange(0, BLOCK)

43 col_mask = col < hidden

vllm/model_executor/layers/quantization/compressed_tensors/schemes/compressed_tensors_wNa8o8.py

+1 -2

@@ -211,8 +211,7 @@ class CompressedTensorsWNA8O8Int(CompressedTensorsScheme):

211

212 def _pack_int_quantized_weight(self, layer: torch.nn.Module) -> None:

213 """Normalize an int-quantized (plain int8) weight to the canonical

214 - ``weight_packed`` int32 + ``weight_shape`` layout the MP kernels expect.

215 - """

214 + ``weight_packed`` int32 + ``weight_shape`` layout the MP kernels expect."""

215 weight = layer.weight

216 out_features, in_features = weight.shape

217 packed = pack_to_int32(weight.data.contiguous(), self.num_bits)

vllm/model_executor/layers/quantization/inc/config_parser.py

+1 -2

@@ -86,8 +86,7 @@ class INCConfigParser:

86

87 def is_explicitly_configured(name: str) -> bool:

88 """Return True if *name* has an explicit entry in extra_config,

89 - either via exact key match or via a regex pattern key.

90 - """

89 + either via exact key match or via a regex pattern key."""

90 if not self._config.extra_config:

91 return False

92 if name in self._config.extra_config:

vllm/model_executor/layers/quantization/inc/schemes/inc_scheme.py

+1 -2

@@ -48,8 +48,7 @@ class INCScheme(ABC):

48 layer_config: "INCLayerConfig",

49 ) -> "FusedMoEMethodBase | None":

50 """Optional. Override if this scheme supports MoE.

51 - Default raises NotImplementedError.

52 - """

51 + Default raises NotImplementedError."""

52 raise NotImplementedError(

53 f"{type(self).__name__} does not support MoE layers. "

54 f"Layer config: {layer_config}"

vllm/model_executor/layers/quantization/modelopt.py

+6 -12

@@ -1988,8 +1988,7 @@ class KNvfp4Static(QuantKeyScheme):

1988

1989 class KNvfp4Dynamic(QuantKeyScheme):

1990 """NVFP4 activation scheme (W4A4). Has a static global input scale on disk;

1991 - the per-group scale is computed at runtime inside the kernel.

1992 - """

1991 + the per-group scale is computed at runtime inside the kernel."""

1992

1993 key = kNvfp4Dynamic

1994

@@ -2026,8 +2025,7 @@ class KNvfp4Dynamic(QuantKeyScheme):

2025

2026 class KFp8StaticTensor(QuantKeyScheme):

2027 """Plain per-tensor static FP8 — bivalent: serves BOTH the weight slot and

2029 - the activation slot (W8A8). One key in both QuantSpec slots.

2030 - """

2028 + the activation slot (W8A8). One key in both QuantSpec slots."""

2029

2030 key = kFp8StaticTensorSym

2031

@@ -2097,8 +2095,7 @@ class KFp8StaticTensor(QuantKeyScheme):

2095

2096 class KFp8StaticChannel(QuantKeyScheme):

2097 """Per-channel static FP8 weight (the 'PcPt' weight). Weight-role only —

2100 - there is no static per-channel *activation* today.

2101 - """

2098 + there is no static per-channel *activation* today."""

2099

2100 key = kFp8StaticTokenSym

2101

@@ -2139,8 +2136,7 @@ class KFp8StaticChannel(QuantKeyScheme):

2136 class KFp8Block128(QuantKeyScheme):

2137 """128x128 block-static FP8 weight ('PbWo'). Weight-role only. ModelOpt

2138 exports the scale 4-D [out_blk,1,in_blk,1]; process squeezes to 2-D.

2142 - No transpose (block kernel keeps [out,in]).

2143 - """

2139 + No transpose (block kernel keeps [out,in])."""

2140

2141 key = kFp8Static128BlockSym

2142

@@ -2196,8 +2192,7 @@ class KFp8Block128(QuantKeyScheme):

2192

2193 class KMxfp8Static(QuantKeyScheme):

2194 """MXFP8 weight: fp8-e4m3 values + per-32-block e8m0 (uint8) scale.

2199 - Weight-role only. process is validate-only plus an idempotency guard.

2200 - """

2195 + Weight-role only. process is validate-only plus an idempotency guard."""

2196

2197 key = kMxfp8Static

2198

@@ -2272,8 +2267,7 @@ class KDynamicNoParam(QuantKeyScheme):

2267 """Dynamic activation with no stored scale (W8A8): quantized at runtime in

2268 the kernel. NOT the same as activation=None (weight-only) — init_fp8 needs a

2269 non-None activation key. Activation-role only. Serves the fp8 per-token, fp8

2275 - per-block, and mxfp8 dynamic activation keys.

2276 - """

2270 + per-block, and mxfp8 dynamic activation keys."""

2271

2272 def create_weights(self, layer, role, ctx, shapes, wl) -> None:

2273 if role is not ACT:

vllm/model_executor/layers/quantization/mxfp4.py

+1 -2

@@ -79,8 +79,7 @@ class Mxfp4Config(QuantizationConfig):

79

80 def _make_moe_method(self, moe: FusedMoEConfig) -> FusedMoEMethodBase:

81 """MoE method for RoutedExperts. Subclasses override to pick a

82 - checkpoint-specific kernel family.

83 - """

82 + checkpoint-specific kernel family."""

83 return Mxfp4MoEMethod(moe)

84

85 def get_quant_method(

vllm/model_executor/layers/quantization/online/base.py

+1 -2

@@ -139,8 +139,7 @@ def _find_matching_targets(

139

140 class OnlineQuantizationConfig(QuantizationConfig):

141 """Model-level config for online quantization (quantize fp16/bf16 weights

142 - during model loading, without requiring a pre-quantized checkpoint).

143 - """

142 + during model loading, without requiring a pre-quantized checkpoint)."""

143

144 def __init__(

145 self,

vllm/model_executor/layers/quantization/online/fp8.py

+6 -12

@@ -118,8 +118,7 @@ def _is_tp_sharded(layer: Module, *, reduces_output_dim: bool = True) -> bool:

118

119 class OnlineLinearBase(LinearMethodBase):

120 """Shared base for online FP8 linear methods. Loads fp16/bf16 checkpoint

121 - weights onto meta device and materializes them just-in-time.

122 - """

121 + weights onto meta device and materializes them just-in-time."""

122

123 uses_meta_device: bool = True

124

@@ -163,8 +162,7 @@ class OnlineLinearBase(LinearMethodBase):

162

163 class Fp8PerTensorOnlineLinearMethod(OnlineLinearBase):

164 """Online tensorwise FP8 linear quantization.

166 - Loads fp16/bf16 weights and quantizes them per-tensor during loading.

167 - """

165 + Loads fp16/bf16 weights and quantizes them per-tensor during loading."""

166

167 def __init__(self):

168 super().__init__()

@@ -265,8 +263,7 @@ class Fp8PerTensorOnlineLinearMethod(OnlineLinearBase):

263

264 class Fp8PerBlockOnlineLinearMethod(OnlineLinearBase):

265 """Online blockwise FP8 linear quantization.

268 - Loads fp16/bf16 weights and quantizes them per-block during loading.

269 - """

266 + Loads fp16/bf16 weights and quantizes them per-block during loading."""

267

268 def __init__(self):

269 super().__init__()

@@ -440,8 +437,7 @@ class Fp8PtpcOnlineLinearMethod(OnlineLinearBase):

437

438 class _Fp8OnlineMoEBase(OnlineMoEMethodBase):

439 """Shared base for online FP8 MoE methods. Loads fp16/bf16 checkpoint

443 - weights onto meta device and materializes them just-in-time.

444 - """

440 + weights onto meta device and materializes them just-in-time."""

441

442 # Declared here for mypy; actual values are set in __init__.

443 fp8_backend: "Fp8MoeBackend"

@@ -562,8 +558,7 @@ class _Fp8OnlineMoEBase(OnlineMoEMethodBase):

558

559 class Fp8PerTensorOnlineMoEMethod(_Fp8OnlineMoEBase):

560 """Online tensorwise FP8 MoE quantization.

565 - Loads fp16/bf16 weights and quantizes them per-tensor during loading.

566 - """

561 + Loads fp16/bf16 weights and quantizes them per-tensor during loading."""

562

563 def __init__(

564 self,

@@ -620,8 +615,7 @@ class Fp8PerTensorOnlineMoEMethod(_Fp8OnlineMoEBase):

615

616 class Fp8PerBlockOnlineMoEMethod(_Fp8OnlineMoEBase):

617 """Online blockwise FP8 MoE quantization.

623 - Loads fp16/bf16 weights and quantizes them per-block during loading.

624 - """

618 + Loads fp16/bf16 weights and quantizes them per-block during loading."""

619

620 def __init__(

621 self,

vllm/model_executor/layers/quantization/utils/fp8_utils.py

+2 -4

@@ -69,8 +69,7 @@ def input_to_float8(

69 x: torch.Tensor, dtype: torch.dtype | None = None

70 ) -> tuple[torch.Tensor, torch.Tensor]:

71 """This function quantizes input values to float8 values "

72 - "with tensor-wise quantization.

73 - """

72 + "with tensor-wise quantization."""

73 dtype = current_platform.fp8_dtype() if dtype is None else dtype

74 finfo = torch.finfo(dtype)

75 min_val, max_val = x.aminmax()

@@ -1195,8 +1194,7 @@ def prepare_fp8_moe_layer_for_deepgemm(

1194

1195 def _maybe_pad_fp8_weight(weight: torch.Tensor) -> torch.Tensor:

1196 """Pad the weight tensor. This is an optimization on ROCm platform, which

1198 - can benefit from tensors located far enough from one another in memory

1199 - """

1197 + can benefit from tensors located far enough from one another in memory"""

1198 if (

1199 envs.VLLM_ROCM_FP8_PADDING

1200 and current_platform.is_rocm()

vllm/model_executor/layers/quantization/utils/humming_utils.py

+2 -4

@@ -297,8 +297,7 @@ def _humming_input_schema_to_quant_key(

297 schema: "HummingInputSchema",

298 ) -> QuantKey | None:

299 """Convert a HummingInputSchema to a QuantKey. Returns None if

300 - the schema represents unquantized (bf16/fp16) inputs.

301 - """

300 + the schema represents unquantized (bf16/fp16) inputs."""

301 if schema.a_dtype is None or schema.a_dtype.num_bits >= 16:

302 return None

303

@@ -427,8 +426,7 @@ def convert_linear_layer_to_humming_standard(

426 ):

427 """Rename/reshape a linear layer's quantized params (the canonical MPLinear

428 layout: ``weight_packed`` int32 + ``weight_scale``) into the parameter names

430 - and layout humming's weight schema expects (``weight`` / ``weight_scale``).

431 - """

429 + and layout humming's weight schema expects (``weight`` / ``weight_scale``)."""

430 for name, checkpoint_name in name_map.items():

431 tensor = getattr(layer, checkpoint_name)

432 delattr(layer, checkpoint_name)

vllm/model_executor/layers/quantization/utils/marlin_utils.py

+3 -6

@@ -245,8 +245,7 @@ def marlin_padded_nk(size_n: int, size_k: int, group_size: int = -1) -> tuple[in

245

246 def marlin_repacked_nk(qweight: torch.Tensor, num_bits: int) -> tuple[int, int]:

247 """Recover the (size_n, size_k) a Marlin weight was repacked with

248 - (including any tile padding) from its packed shape.

249 - """

248 + (including any tile padding) from its packed shape."""

249 pack_factor = 32 // num_bits

250 size_k = qweight.size(0) * GPTQ_MARLIN_TILE

251 size_n = qweight.size(1) * pack_factor // GPTQ_MARLIN_TILE

@@ -257,8 +256,7 @@ def marlin_pad_qweight(

256 qweight: torch.Tensor, size_n: int, size_k: int, padded_n: int, padded_k: int

257 ) -> torch.Tensor:

258 """Zero-pad a GPTQ-layout packed weight (size_k / pack, size_n) for

260 - gptq_marlin_repack.

261 - """

259 + gptq_marlin_repack."""

260 if (padded_n, padded_k) == (size_n, size_k):

261 return qweight

262 pack_factor = size_k // qweight.size(0)

@@ -276,8 +274,7 @@ def marlin_pad_scales(

274 group_size: int,

275 ) -> torch.Tensor:

276 """Zero-pad weight scales (num_groups, size_n); call before

279 - marlin_permute_scales and pass the padded extents to it.

280 - """

277 + marlin_permute_scales and pass the padded extents to it."""

278 if (padded_n, padded_k) == (size_n, size_k):

279 return scales

280 pad_rows = padded_k // group_size - scales.size(0) if group_size > 0 else 0

vllm/model_executor/layers/quantization/utils/marlin_utils_fp4.py

+3 -6

@@ -41,8 +41,7 @@ def _nvfp4_compute_scale_factor(

41 ) -> float:

42 """Compute the power-of-2 scale_factor needed so that all non-zero

43 values in marlin_scales * 2^7 are >= 2 after rescaling.

44 - Returns a Python float (power of 2, >= 1.0).

45 - """

44 + Returns a Python float (power of 2, >= 1.0)."""

45 # Since half has a smaller dynamic range compared to bfloat16,

46 # no rescaling is applied here if active dtype is half.

47 if a_dtype is not None and a_dtype == torch.half:

@@ -376,8 +375,7 @@ def prepare_nvfp4_moe_layer_for_marlin(

375

376 def pad_w13(x: torch.Tensor) -> torch.Tensor:

377 """Zero-pad each gate/up shard of a (E, num_shards * N, cols)

379 - tensor to padded_N rows.

380 - """

378 + tensor to padded_N rows."""

379 if padded_N == N:

380 return x

381 x = x.view(E, num_shards, N, x.size(-1))

@@ -386,8 +384,7 @@ def prepare_nvfp4_moe_layer_for_marlin(

384

385 def pad_w2(x: torch.Tensor, packing: int) -> torch.Tensor:

386 """Zero-pad the packed N (last) dim of a (E, K, N / packing)

389 - tensor.

390 - """

387 + tensor."""

388 if padded_N == N:

389 return x

390 return torch.nn.functional.pad(x, (0, (padded_N - N) // packing))

vllm/model_executor/layers/quantization/utils/marlin_utils_fp8.py

+1 -2

@@ -217,8 +217,7 @@ def prepare_fp8_layer_for_marlin(

217

218 def _moe_pad_shard_rows(x: torch.Tensor, n: int, padded_n: int) -> torch.Tensor:

219 """Zero-pad each gate/up shard of a ``(E, 2 * n, ...)`` tensor to padded_n

220 - rows. FP8 zero decodes to 0.0, so the padded rows contribute nothing.

221 - """

220 + rows. FP8 zero decodes to 0.0, so the padded rows contribute nothing."""

221 if padded_n == n:

222 return x

223 e = x.size(0)

vllm/model_executor/layers/rotary_embedding/ntk_scaling_rope.py

+1 -2

@@ -9,8 +9,7 @@ from .base import RotaryEmbedding

9

10 class NTKScalingRotaryEmbedding(RotaryEmbedding):

11 """RotaryEmbedding extended with fixed and mixed NTK scaling.

12 - https://kexue.fm/archives/9706

13 - """

12 + https://kexue.fm/archives/9706"""

13

14 def __init__(

15 self,

vllm/model_executor/layers/sparse_attn_indexer.py

+2 -4

@@ -285,8 +285,7 @@ def _gather_workspace_shapes(

285 """Return ((values_shape, values_dtype), (scales_shape, scales_dtype)) for

286 the K-gather workspace. FP8 path: (T, head_dim) fp8 + (T, 4) uint8 fp32

287 scales. MXFP4 path: (T, head_dim // 2) uint8 packed mxfp4 +

288 - (T, head_dim // MXFP4_BLOCK_SIZE) uint8 ue8m0 scales.

289 - """

288 + (T, head_dim // MXFP4_BLOCK_SIZE) uint8 ue8m0 scales."""

289 if use_fp4_cache:

290 return (

291 ((total_seq_lens, head_dim // 2), torch.uint8),

@@ -304,8 +303,7 @@ def kv_cache_as_quant_view(

303 use_fp4_cache: bool,

304 ) -> torch.Tensor:

305 """4D ``[num_blocks, block_size, 1, head_width]`` view expected by

307 - DeepGEMM, from the 3D indexer kv-cache allocation.

308 - """

306 + DeepGEMM, from the 3D indexer kv-cache allocation."""

307 if use_fp4_cache:

308 assert kv_cache.ndim == 3 and kv_cache.dtype == torch.uint8

309 num_blocks, block_size, _ = kv_cache.shape

vllm/model_executor/layers/sparse_mqa_indexer.py

+2 -4

@@ -41,8 +41,7 @@ def _prefill_k_workspaces(

41 total_seq_lens: int, head_dim: int

42 ) -> tuple[torch.Tensor, torch.Tensor]:

43 """The packed MXFP4 K-gather workspace ``(values, scales)``, shared with

44 - the dense indexer layers of the same model.

45 - """

44 + the dense indexer layers of the same model."""

45 values_spec, scales_spec = _gather_workspace_shapes(

46 total_seq_lens, head_dim, current_platform.fp8_dtype(), use_fp4_cache=True

47 )

@@ -120,8 +119,7 @@ class SparseMQAIndexer(nn.Module):

119

120 def _reserve_workspaces(self, device: torch.device) -> None:

121 """Profiling run: claim the K-gather workspace and the peak sparse

123 - logits allocation so the memory estimate covers them.

124 - """

122 + logits allocation so the memory estimate covers them."""

123 _prefill_k_workspaces(self.max_total_seq_len, self.head_dim)

124 max_logits_bytes = envs.VLLM_SPARSE_INDEXER_MAX_LOGITS_MB * 1024 * 1024

125 torch.empty(max_logits_bytes, dtype=torch.uint8, device=device)

vllm/model_executor/layers/vocab_parallel_embedding.py

+1 -2

@@ -376,8 +376,7 @@ class VocabParallelEmbedding(PluggableLayer):

376 ) -> VocabParallelEmbeddingShardIndices:

377 """Get start and end indices for vocab parallel embedding, following the

378 layout outlined in the class docstring, based on the given tp_rank and

379 - tp_size.

380 - """

379 + tp_size."""

380 num_added_embeddings_padded = vocab_size_padded - org_vocab_size_padded

381 padded_org_vocab_start_index, padded_org_vocab_end_index = (

382 vocab_range_from_global_vocab_size(org_vocab_size_padded, tp_rank, tp_size)

vllm/model_executor/model_loader/base_loader.py

+1 -2

@@ -36,8 +36,7 @@ class BaseModelLoader(ABC):

36 @abstractmethod

37 def load_weights(self, model: nn.Module, model_config: ModelConfig) -> None:

38 """Load weights into a model. This standalone API allows

39 - inplace weights loading for an already-initialized model

40 - """

39 + inplace weights loading for an already-initialized model"""

40 raise NotImplementedError

41

42 def create_model(

vllm/model_executor/model_loader/default_loader.py

+1 -2

@@ -135,8 +135,7 @@ class DefaultModelLoader(BaseModelLoader):

135 ) -> tuple[str, list[str], bool]:

136 """Prepare weights for the model.

137

138 - If the model is not local, it will be downloaded.

139 - """

138 + If the model is not local, it will be downloaded."""

139 model_name_or_path = (

140 maybe_download_from_modelscope(model_name_or_path, revision)

141 or model_name_or_path

vllm/model_executor/model_loader/ep_weight_filter.py

+2 -4

@@ -23,8 +23,7 @@ def parse_expert_id(weight_name: str) -> int | None:

23

24 Returns ``None`` for dense weights (attention, layernorm, embedding),

25 shared experts, and 3D fused-expert tensors where all experts are stored

26 - in a single tensor without a numeric expert id in the name.

27 - """

26 + in a single tensor without a numeric expert id in the name."""

27 m = _EXPERT_ID_RE.search(weight_name)

28 return int(m.group(1)) if m else None

29

@@ -71,8 +70,7 @@ def should_skip_weight(

70 local_expert_ids: set[int] | None,

71 ) -> bool:

72 """Return ``True`` if *weight_name* is an expert weight that does not

74 - belong to the local rank and should be skipped during loading.

75 - """

73 + belong to the local rank and should be skipped during loading."""

74 if local_expert_ids is None:

75 return False

76 eid = parse_expert_id(weight_name)

vllm/model_executor/model_loader/reload/layerwise.py

+2 -4

@@ -308,8 +308,7 @@ def _reload_attention_scales(layer: torch.nn.Module, info: LayerReloadingInfo) -

308

309 Assumes dtype/shapes of attention tensors do not change during

310 processing, since we use .data.copy_() to preserve kernel tensor

311 - references.

312 - """

311 + references."""

312 quant_method = getattr(layer, "quant_method", None)

313 if quant_method is not None:

314 # Re-create scale Parameters with sentinel values so unloaded scales

@@ -388,8 +387,7 @@ def _get_weight_loader(tensor: torch.Tensor):

387

388 def _copy_and_restore_kernel_tensors(layer: torch.nn.Module, info: LayerReloadingInfo):

389 """Copy processed values into original kernel tensor storage and restore

391 - kernel tensor references on the layer. Preserves cudagraph references.

392 - """

390 + kernel tensor references on the layer. Preserves cudagraph references."""

391 assert info.kernel_tensors is not None

392 parameters, buffers = info.kernel_tensors

393 non_persistent = info.kernel_non_persistent_buffers

vllm/model_executor/model_loader/runai_streamer_loader.py

+1 -2

@@ -81,8 +81,7 @@ class RunaiModelStreamerLoader(BaseModelLoader):

81 ) -> list[str]:

82 """Prepare weights for the model.

83

84 - If the model is not local, it will be downloaded.

85 - """

84 + If the model is not local, it will be downloaded."""

85 is_object_storage_path = is_runai_obj_uri(model_name_or_path)

86 is_local = os.path.isdir(model_name_or_path)

87 safetensors_pattern = "*.safetensors"

vllm/model_executor/model_loader/tensorizer.py

+1 -2

@@ -497,8 +497,7 @@ def _check_tensors_on_meta_device(model: nn.Module) -> None:

497

498 def _resize_lora_embeddings(model: nn.Module):

499 """Modify LoRA embedding layers to use bigger tensors

500 - to allow for adapter added tokens.

501 - """

500 + to allow for adapter added tokens."""

501 for child in model.modules():

502 if (

503 isinstance(child, VocabParallelEmbedding)

vllm/model_executor/model_loader/tensorizer_loader.py

+1 -2

@@ -105,8 +105,7 @@ class TensorizerLoader(BaseModelLoader):

105

106 Expects a vLLM-tensorized model. See the

107 examples/features/tensorize_vllm_model.py example script

108 - for serializing vLLM models.

109 - """

108 + for serializing vLLM models."""

109 if is_vllm_tensorized(self.tensorizer_config):

110 tensorizer_config = self._patch_tensorizer_config(model_config)

111 deserialize_tensorizer_model(model, tensorizer_config)

vllm/model_executor/model_loader/weight_cache/ipc_loader.py

+1 -2

@@ -1,8 +1,7 @@

1 # SPDX-License-Identifier: Apache-2.0

2 # SPDX-FileCopyrightText: Copyright contributors to the vLLM project

3 """IPC model loader: maps post-quantized weights from a local weight cache

4 -daemon via CUDA IPC instead of loading from disk.

5 -"""

4 +daemon via CUDA IPC instead of loading from disk."""

5

6 import dataclasses

7 import socket

vllm/model_executor/model_loader/weight_tying.py

+1 -2

@@ -78,8 +78,7 @@ def maybe_retie_word_embeddings(model: nn.Module, model_config: ModelConfig) ->

78

79 Checkpoints produced by quantization or fine-tuning tooling often keep a redundant

80 copy of the tied `lm_head`. Sharing the storage again reclaims the memory it would

81 - otherwise cost.

82 - """

81 + otherwise cost."""

82 if not model_config.word_embeddings_untied_by_checkpoint:

83 return

84 if (untied := _get_untied_lm_head(model)) is None:

vllm/model_executor/model_loader/weight_utils.py

+3 -6

@@ -144,8 +144,7 @@ def atomic_writer(

144

145 def _natural_sort_key(filepath: str) -> list:

146 """Natural sort key for filenames with numeric components, such as

147 - model-00001-of-00005.safetensors -> ['model-', 1, '-of-', 5, '.safetensors']

148 - """

147 + model-00001-of-00005.safetensors -> ['model-', 1, '-of-', 5, '.safetensors']"""

148 return [

149 int(s) if s.isdigit() else s

150 for s in re.split(r"(\d+)", os.path.basename(filepath))

@@ -162,8 +161,7 @@ def maybe_download_from_modelscope(

161 """Download model from ModelScope hub if VLLM_USE_MODELSCOPE is True.

162

163 Returns the path to the downloaded model, or None if the model is not

165 - downloaded from ModelScope.

166 - """

164 + downloaded from ModelScope."""

165 if envs.VLLM_USE_MODELSCOPE:

166 # download model from ModelScope hub,

167 # lazy import so that modelscope is not required for normal use.

@@ -1148,8 +1146,7 @@ def instanttensor_weights_iterator(

1146 use_tqdm_on_load: bool,

1147 ) -> Generator[tuple[str, torch.Tensor], None, None]:

1148 """Iterate over the weights in the model safetensor files

1151 - using instanttensor library.

1152 - """

1149 + using instanttensor library."""

1150 try:

1151 import instanttensor

1152 except ImportError as e:

vllm/model_executor/models/arcee.py

+5 -10

@@ -47,8 +47,7 @@ from .utils import (

47

48 class ArceeMLP(nn.Module):

49 """Feed-forward layer for Arcee using ReLU^2 activation

50 - (no gating as in LLaMA).

51 - """

50 + (no gating as in LLaMA)."""

51

52 def __init__(

53 self,

@@ -96,8 +95,7 @@ class ArceeMLP(nn.Module):

95

96 class ArceeDecoderLayer(nn.Module):

97 """Transformer decoder block for Arcee, with self-attention and

99 - ReLU^2 MLP.

100 - """

98 + ReLU^2 MLP."""

99

100 def __init__(

101 self,

@@ -177,8 +175,7 @@ class ArceeDecoderLayer(nn.Module):

175 @support_torch_compile

176 class ArceeModel(nn.Module, EagleModelMixin):

177 """The transformer model backbone for Arcee (embedding layer + stacked

180 - decoder blocks + final norm).

181 - """

178 + decoder blocks + final norm)."""

179

180 def __init__(

181 self,

@@ -280,8 +277,7 @@ class ArceeForCausalLM(

277 nn.Module, SupportsLoRA, SupportsPP, SupportsEagle, SupportsEagle3

278 ):

279 """Arcee Model for causal language modeling, integrated with vLLM

283 - runtime.

284 - """

280 + runtime."""

281

282 hf_to_vllm_mapper = WeightsMapper(

283 orig_to_new_stacked={

@@ -362,8 +358,7 @@ class ArceeForCausalLM(

358

359 def load_weights(self, weights: Iterable[tuple[str, torch.Tensor]]) -> set[str]:

360 """Load weights into the model (delegates to inner model and handles

365 - tied embeddings).

366 - """

361 + tied embeddings)."""

362 loader = AutoWeightsLoader(self)

363 # AutoWeightLoader handles weight name remapping, including fusing

364 # separate q_proj, k_proj, v_proj into qkv_proj

vllm/model_executor/models/blip.py

+1 -2

@@ -1,8 +1,7 @@

1 # SPDX-License-Identifier: Apache-2.0

2 # SPDX-FileCopyrightText: Copyright contributors to the vLLM project

3 """Minimal implementation of BlipVisionModel intended to be only used

4 -within a vision language model.

5 -"""

4 +within a vision language model."""

5

6 from collections.abc import Iterable

7

vllm/model_executor/models/cohere2_moe.py

+1 -2

@@ -96,8 +96,7 @@ class RMSNorm(nn.Module):

96

97 def select_norm_impl(config: CohereConfig) -> tuple[type[nn.Module], float]:

98 """Returns (norm_class, eps). Uses RMSNorm when config.rms_norm_eps is set,

99 - otherwise falls back to LayerNorm with config.layer_norm_eps.

100 - """

99 + otherwise falls back to LayerNorm with config.layer_norm_eps."""

100 rms_eps = getattr(config, "rms_norm_eps", None)

101 if rms_eps is not None:

102 return RMSNorm, rms_eps

vllm/model_executor/models/cohere_asr.py

+1 -2

@@ -667,8 +667,7 @@ class ConvSubsampling(nn.Module):

667 repeat_num: int = 1,

668 ) -> torch.Tensor:

669 """Calculates the output length of a Tensor passed

670 - through a convolution or max pooling layer

671 - """

670 + through a convolution or max pooling layer"""

671 add_pad: float = all_paddings - kernel_size

672 one: float = 1.0

673 for i in range(repeat_num):

vllm/model_executor/models/deepencoder.py

+1 -2

@@ -363,8 +363,7 @@ class ImageEncoderViT(nn.Module):

363

364 class Block(nn.Module):

365 """Transformer blocks with support of window attention and residual propagation

366 - blocks

367 - """

366 + blocks"""

367

368 def __init__(

369 self,

vllm/model_executor/models/deepseek_ocr.py

+1 -2

@@ -139,8 +139,7 @@ class NoRepeatNGramLogitsProcessor:

139

140 class NGramPerReqLogitsProcessor(AdapterLogitsProcessor):

141 """Example of overriding the wrapper class `__init__()` in order to utilize

142 - info about the device type

143 - """

142 + info about the device type"""

143

144 @classmethod

145 def validate_params(cls, params: SamplingParams):

vllm/model_executor/models/diffusion_gemma.py

+1 -2

@@ -511,8 +511,7 @@ def _compiled_sample_step(

511 accept/renoise → convergence, all as vectorized PyTorch ops.

512

513 Returns the temperature-scaled logits ``[num_decode, CL, vocab]`` so the

514 - caller can compute logprobs outside the compiled region.

515 - """

514 + caller can compute logprobs outside the compiled region."""

515 num_decode = decode_slots.shape[0]

516 device = decode_slots.device

517

vllm/model_executor/models/gemma3n.py

+1 -2

@@ -814,8 +814,7 @@ class Gemma3nCrossDecoder(nn.Module):

814 def _kv_sharing_weights_mapper(config: Gemma3nTextConfig) -> WeightsMapper:

815 """KV-shared layers only have q_proj, so qkv_proj packing applies to the

816 other layers. Original checkpoints still ship K/V tensors for the shared

817 - layers (fine-tuned ones omit them); those are dropped.

818 - """

817 + layers (fine-tuned ones omit them); those are dropped."""

818 first_kv_shared_layer_idx = config.num_hidden_layers - config.num_kv_shared_layers

819 return WeightsMapper(

820 orig_to_new_substr={

vllm/model_executor/models/gemma3n_mm.py

+1 -2

@@ -499,8 +499,7 @@ class Gemma3nMultiModalProcessor(BaseMultiModalProcessor[Gemma3nProcessingInfo])

499

500 class Gemma3nMultimodalEmbedder(nn.Module):

501 """Embeds token ids or soft tokens for multimodal content into language

502 - model space.

503 - """

502 + model space."""

503

504 def __init__(

505 self,

vllm/model_executor/models/glm4_1v.py

+1 -2

@@ -25,8 +25,7 @@

25 # See the License for the specific language governing permissions and

26 # limitations under the License.

27 """Inference-only GLM-4.1V & GLM-4.6V-Flash, AutoGLM-Phone-9B model

28 -compatible with HuggingFace weights.

29 -"""

28 +compatible with HuggingFace weights."""

29

30 import math

31 from collections.abc import Callable, Hashable, Iterable, Iterator, Mapping, Sequence

vllm/model_executor/models/glm4_moe.py

+1 -2

@@ -22,8 +22,7 @@

22 # See the License for the specific language governing permissions and

23 # limitations under the License.

24 """Inference-only GLM-4.5, GLM-4.6, GLM-4.7 model

25 -compatible with HuggingFace weights.

26 -"""

25 +compatible with HuggingFace weights."""

26

27 from collections.abc import Iterable

28 from itertools import islice

vllm/model_executor/models/glm4_moe_mtp.py

+1 -2

@@ -22,8 +22,7 @@

22 # See the License for the specific language governing permissions and

23 # limitations under the License.

24 """Inference-only GLM-4.5, GLM-4.6, GLM-4.7 MTP

25 -model compatible with HuggingFace weights.

26 -"""

25 +model compatible with HuggingFace weights."""

26

27 import typing

28 from collections.abc import Callable, Iterable

vllm/model_executor/models/granite_speech.py

+1 -2

@@ -462,8 +462,7 @@ class GraniteSpeechConformerConvModule(nn.Module):

462

463 class GraniteSpeechConformerBlock(nn.Module):

464 """Conformer block, consisting largely of linear layers,

465 - attention, and convolutional layers.

466 - """

465 + attention, and convolutional layers."""

466

467 def __init__(self, config: PretrainedConfig, prefix: str = ""):

468 super().__init__()

vllm/model_executor/models/interfaces.py

+5 -10

@@ -892,8 +892,7 @@ def has_inner_state(

892 @runtime_checkable

893 class IsAttentionFree(Protocol):

894 """The interface required for all models like Mamba that lack attention,

895 - but do have state whose size is constant wrt the number of tokens.

896 - """

895 + but do have state whose size is constant wrt the number of tokens."""

896

897 is_attention_free: ClassVar[Literal[True]] = True

898 """

@@ -921,8 +920,7 @@ def is_attention_free(

920 class IsHybrid(Protocol):

921 """The interface required for all models like Jamba that have both

922 attention and mamba blocks, indicates that

924 - hf_config has 'layers_block_type'

925 - """

923 + hf_config has 'layers_block_type'"""

924

925 is_hybrid: ClassVar[Literal[True]] = True

926 """

@@ -1338,8 +1336,7 @@ class SupportsTranscription(Protocol):

1336 ) -> "PromptType":

1337 """Get the prompt for the ASR model.

1338 The model has control over the construction, as long as it

1341 - returns a valid PromptType.

1342 - """

1339 + returns a valid PromptType."""

1340 ...

1341

1342 @classmethod

@@ -1631,8 +1628,7 @@ class EagleModelMixin:

1628 @runtime_checkable

1629 class SupportsEagle(SupportsEagleBase, Protocol):

1630 """The interface required for models that support

1634 - EAGLE-1 and EAGLE-2 speculative decoding.

1635 - """

1631 + EAGLE-1 and EAGLE-2 speculative decoding."""

1632

1633 supports_eagle: ClassVar[Literal[True]] = True

1634 """

@@ -1662,8 +1658,7 @@ def supports_eagle(

1658 @runtime_checkable

1659 class SupportsEagle3(SupportsEagleBase, Protocol):

1660 """The interface required for models that support

1665 - EAGLE-3 speculative decoding.

1666 - """

1661 + EAGLE-3 speculative decoding."""

1662

1663 supports_eagle3: ClassVar[Literal[True]] = True

1664 """

vllm/model_executor/models/keye.py

+1 -2

@@ -345,8 +345,7 @@ def apply_rotary_pos_emb_flashatt(

345

346 class KeyeSiglipAttention(nn.Module):

347 """Multi-headed attention from 'Attention Is All You

348 - Need' paper.

349 - """

348 + Need' paper."""

349

350 def __init__(

351 self,

vllm/model_executor/models/lfm2_siglip2.py

+1 -2

@@ -1,8 +1,7 @@

1 # SPDX-License-Identifier: Apache-2.0

2 # SPDX-FileCopyrightText: Copyright contributors to the vLLM project

3 """Implementation of Siglip2VisionModel intended to be only used

4 -within a vision language model.

5 -"""

4 +within a vision language model."""

5

6 from collections.abc import Iterable

7

vllm/model_executor/models/medusa.py

+1 -2

@@ -51,8 +51,7 @@ class Medusa(nn.Module):

51 truncated_vocab_size < vocab_size. To use this technique, one has to find

52 the top-k most frequent tokens in target dataset and add that as a tensor

53 in the draft checkpoint (using key token_map). Also, the draft config

54 - needs to have truncated_vocab_size (=k) as an attribute.

55 - """

54 + needs to have truncated_vocab_size (=k) as an attribute."""

55

56 def __init__(self, *, vllm_config: VllmConfig, prefix: str = "") -> None:

57 speculative_config = vllm_config.speculative_config

vllm/model_executor/models/minicpmv4_6.py

+1 -2

@@ -858,8 +858,7 @@ class MiniCPMV4_6ViTWindowAttentionMerger(nn.Module):

858

859 class MiniCPMV4_6DownsampleMLP(nn.Module):

860 """Match HF (transformers v5.7+) parameter naming: pre_norm/linear_1/

861 - act/linear_2 (instead of pre_norm + Sequential(mlp.0/mlp.2)).

862 - """

861 + act/linear_2 (instead of pre_norm + Sequential(mlp.0/mlp.2))."""

862

863 def __init__(

864 self,

vllm/model_executor/models/mllama4.py

+1 -2

@@ -1090,8 +1090,7 @@ class Llama4ForConditionalGeneration(

1090

1091 def _rename_weight_for_modelopt_checkpoint(self, name: str) -> str:

1092 """Rename weights from ModelOpt llama4 fp8 checkpoints to vLLM

1093 - format.

1094 - """

1093 + format."""

1094 if name.startswith("model.") or name.startswith("language_model.model."):

1095 renamed = (

1096 name.replace("model.", "language_model.model.", 1)

vllm/model_executor/models/muse_glimmer.py

+3 -6

@@ -114,8 +114,7 @@ logger = init_logger(__name__)

114 def _text_config(config):

115 """MuseGlimmer checkpoints may nest the text config under ``text_config``

116 (multimodal ``MuseGlimmerConfig``) or expose it directly

117 - (``MuseGlimmerTextConfig``).

118 - """

117 + (``MuseGlimmerTextConfig``)."""

118 return getattr(config, "text_config", config)

119

120

@@ -489,8 +488,7 @@ class MuseGlimmerMultiModalProcessor(

488 def _muse_glimmer_use_qk_norm(config) -> bool:

489 """Whether QK-norm is applied. MuseGlimmer ALWAYS applies QK-norm; the modular HF

490 ``text_config`` schema simply omits ``use_qk_norm`` (reads as ``None``).

492 - Treat a missing/None flag as True — only an explicit ``False`` disables it.

493 - """

491 + Treat a missing/None flag as True — only an explicit ``False`` disables it."""

492 val = getattr(config, "use_qk_norm", None)

493 return True if val is None else bool(val)

494

@@ -498,8 +496,7 @@ def _muse_glimmer_use_qk_norm(config) -> bool:

496 def _muse_glimmer_use_attn_output_gate(config) -> bool:

497 """Whether the per-head sigmoid attention output gate is applied. MuseGlimmer ALWAYS

498 applies it; the modular HF ``text_config`` omits ``use_attn_output_gate``

501 - (reads as ``None``). Missing/None -> True; only explicit ``False`` disables.

502 - """

499 + (reads as ``None``). Missing/None -> True; only explicit ``False`` disables."""

500 val = getattr(config, "use_attn_output_gate", None)

501 return True if val is None else bool(val)

502

vllm/model_executor/models/nano_nemotron_vl.py

+1 -2

@@ -1188,8 +1188,7 @@ class NemotronH_Nano_VL_V2(

1188 self, video_input: NanoNemotronVLVideoPixelInputs

1189 ) -> tuple[torch.Tensor, ...]:

1190 """Process video input and create final embeddings with video content

1191 - and indicator tokens.

1192 - """

1191 + and indicator tokens."""

1192 T = self.video_temporal_patch_size

1193

1194 if T > 1:

vllm/model_executor/models/phi4mm_utils.py

+1 -2

@@ -1462,8 +1462,7 @@ class NemoConvSubsampling(torch.nn.Module):

1462

1463 def conv_split_by_channel(self, x: Tensor) -> Tensor:

1464 """For dw convs, tries to split input by time, run conv and concat

1465 - results

1466 - """

1465 + results"""

1466 x = self.conv[0](x) # full conv2D

1467 x = self.conv[1](x) # activation

1468

vllm/model_executor/models/qwen3_dflash.py

+4 -8

@@ -69,8 +69,7 @@ def _dflash_layer_causal(config: Qwen3Config, layer_idx: int) -> bool:

69

70 def dflash_has_any_non_causal(config: Qwen3Config) -> bool:

71 """Whether the draft needs a non-causal-capable backend, resolved from config

72 - (config mirror of the model's ``get_draft_attn_causal``, usable pre-build).

73 - """

72 + (config mirror of the model's ``get_draft_attn_causal``, usable pre-build)."""

73 return not all(

74 _dflash_layer_causal(config, i) for i in range(config.num_hidden_layers)

75 )

@@ -155,8 +154,7 @@ class DFlashQwen3Attention(nn.Module):

154

155 Context KVs are pre-inserted into the KV cache before the forward pass.

156 This layer handles only query tokens via standard attention.

158 - Adapted from Qwen3Attention.

159 - """

157 + Adapted from Qwen3Attention."""

158

159 def __init__(

160 self,

@@ -250,8 +248,7 @@ class DFlashQwen3Attention(nn.Module):

248 """DFlash attention assumes that the KV cache is already populated

249 with the context K/V from the target model's hidden states. This forward op

250 computes attention for the query tokens only.

253 - See also: precompute_and_store_context_kv

254 - """

251 + See also: precompute_and_store_context_kv"""

252 qkv, _ = self.qkv_proj(hidden_states)

253 q, k, v = qkv.split([self.q_size, self.kv_size, self.kv_size], dim=-1)

254

@@ -743,8 +740,7 @@ class DFlashQwen3ForCausalLM(Qwen3ForCausalLM):

740

741 def get_draft_attn_causal(self) -> list[bool]:

742 """Per-layer attention causality, aligned with

746 - get_draft_kv_cache_layer_names.

747 - """

743 + get_draft_kv_cache_layer_names."""

744 return [layer.self_attn.causal for layer in self.model.layers]

745

746 def compute_logits(

vllm/model_executor/models/qwen3_vl.py

+1 -2

@@ -1321,8 +1321,7 @@ class Qwen3VLMultiModalProcessor(BaseMultiModalProcessor[Qwen3VLProcessingInfo])

1321 def _expands_only_video_token(hf_processor: ProcessorMixin) -> bool:

1322 """Transformers>=5.10 processors override `replace_video_token`

1323 to expand only the bare video token, keeping the prompt's outer

1324 - `<|vision_start|>`/`<|vision_end|>` markers.

1325 - """

1324 + `<|vision_start|>`/`<|vision_end|>` markers."""

1325 mixin_impl = getattr(ProcessorMixin, "replace_video_token", None)

1326 proc_impl = getattr(type(hf_processor), "replace_video_token", None)

1327 return proc_impl is not None and proc_impl is not mixin_impl

vllm/model_executor/models/radio.py

+2 -4

@@ -599,8 +599,7 @@ class RadioInternVisionModel(nn.Module):

599 self, imgs_sizes: list[tuple[int, int]], device: torch.device

600 ) -> MaskMetadata:

601 """Build mask metadata from image pixel sizes. Adds num_skip to each

602 - sequence length (cls/register tokens) to match patch generator output.

603 - """

602 + sequence length (cls/register tokens) to match patch generator output."""

603 patch_size = self.patch_generator.patch_size

604 num_skip = self.patch_generator.num_skip

605

@@ -613,8 +612,7 @@ class RadioInternVisionModel(nn.Module):

612 ) -> MaskMetadata:

613 """Build mask metadata from actual sequence lengths (already including

614 cls/register tokens, i.e. patch_count + num_skip per item).

616 - Use inter_image_mask_metadata() when you only have imgs_sizes.

617 - """

615 + Use inter_image_mask_metadata() when you only have imgs_sizes."""

616 assert len(seq_lens) > 0

617 cu_seqlens = torch.tensor(

618 list(accumulate(seq_lens, initial=0)), dtype=torch.int32, device=device

vllm/model_executor/models/siglip2navit.py

+1 -2

@@ -1,8 +1,7 @@

1 # SPDX-License-Identifier: Apache-2.0

2 # SPDX-FileCopyrightText: Copyright contributors to the vLLM project

3 """Implementation of SiglipVisionModel intended to be only used

4 -within a vision language model.

5 -"""

4 +within a vision language model."""

5

6 from collections.abc import Iterable

7

vllm/model_executor/models/stablelm.py

+1 -2

@@ -20,8 +20,7 @@

20 # https://huggingface.co/stabilityai/stablelm-3b-4e1t/blob/main/modeling_stablelm_epoch.py

21 # https://huggingface.co/stabilityai/stablelm-3b-4e1t/blob/main/config.json

22 """Inference-only StableLM (https://github.com/Stability-AI/StableLM)

23 -model compatible with HuggingFace weights.

24 -"""

23 +model compatible with HuggingFace weights."""

24

25 from collections.abc import Iterable

26 from itertools import islice

vllm/model_executor/models/transformers/base.py

+2 -4

@@ -207,8 +207,7 @@ class Base(

207 def _mark_model_components(self, vllm_config: "VllmConfig"):

208 """Mark language model and tower submodules as `self.model` is created.

209

210 - Nothing to do in `Base`, `MultiModalMixin` will override.

211 - """

210 + Nothing to do in `Base`, `MultiModalMixin` will override."""

211 yield

212

213 @cached_property

@@ -281,8 +280,7 @@ class Base(

280

281 def _decorate_for_torch_compile(self):

282 """Decorate the model's decoder class to indicate to vLLM that it

284 - supports torch compile if `can_enable_torch_compile` is True.

285 - """

283 + supports torch compile if `can_enable_torch_compile` is True."""

284 self._decorate_cls_for_torch_compile(

285 cls=self._pre_trained_model_classes.decoder,

286 # Applied to a PreTrainedModel so the batch dimension will exist

vllm/model_executor/models/transformers/fusers/base.py

+9 -18

@@ -55,20 +55,17 @@ class BaseFuser(ABC):

55 self, module: nn.Module, prefix: str, vllm_config: "VllmConfig"

56 ) -> nn.Module:

57 """Apply the fusion to an already-validated `module`, returning the

58 - module to install in its place (mutated in place, or freshly built).

59 - """

58 + module to install in its place (mutated in place, or freshly built)."""

59

60 def orig_to_new_stacked(self, prefix: str) -> dict[str, tuple[str, ShardId]]:

61 """`WeightsMapper.orig_to_new_stacked` entries this fuser contributes

63 - (none unless it stacks weights).

64 - """

62 + (none unless it stacks weights)."""

63 return {}

64

65 @property

66 def packed_modules_mapping(self) -> dict[str, list[str]]:

67 """`packed_modules_mapping` entries this fuser contributes (none unless

70 - it stacks weights).

71 - """

68 + it stacks weights)."""

69 return {}

70

71

@@ -122,8 +119,7 @@ class RewriteFuser(BaseFuser):

119 ) -> nn.Module:

120 """Fuse an already-validated `module` in place (see `Fusers.__getitem__`).

121

125 - Builds the merged submodule and binds the compiled forward.

126 - """

122 + Builds the merged submodule and binds the compiled forward."""

123 self.update_attrs(module, prefix, vllm_config)

124 module.forward = types.MethodType(self.fused_forward, module)

125 return module

@@ -132,8 +128,7 @@ class RewriteFuser(BaseFuser):

128 @dataclass

129 class StackedFuser(RewriteFuser):

130 """A fuser that merges sibling projections into one stacked linear and

135 - rewrites the forward to call it.

136 - """

131 + rewrites the forward to call it."""

132

133 merged_name: ClassVar[str]

134 """Attribute name of the merged module created by `update_attrs`."""

@@ -152,16 +147,14 @@ class StackedFuser(RewriteFuser):

147 def shards(self) -> list[tuple[str, ShardId]]:

148 """Each projection's original name and its shard id in the merged module.

149

155 - Source for both `orig_to_new_stacked` and `packed_modules_mapping`.

156 - """

150 + Source for both `orig_to_new_stacked` and `packed_modules_mapping`."""

151

152 def orig_to_new_stacked(self, prefix: str) -> dict[str, tuple[str, ShardId]]:

153 """`WeightsMapper.orig_to_new_stacked` entries for one fused instance.

154

155 Maps each checkpoint name to `(merged_name, shard_id)`, keyed by qualname

156 so only this exact layer is remapped, never a same-named projection

163 - elsewhere (e.g. an unfused MoE expert's `gate_proj`).

164 - """

157 + elsewhere (e.g. an unfused MoE expert's `gate_proj`)."""

158 merged = maybe_prefix(prefix, self.merged_name)

159 return {

160 maybe_prefix(prefix, name): (merged, shard) for name, shard in self.shards

@@ -170,8 +163,7 @@ class StackedFuser(RewriteFuser):

163 @property

164 def packed_modules_mapping(self) -> dict[str, list[str]]:

165 """`{merged_name: [projection names]}` so quantization can unpack the

173 - fused layer into its per-shard configs.

174 - """

166 + fused layer into its per-shard configs."""

167 return {self.merged_name: [name for name, _ in self.shards]}

168

169 def _unguarded_calls(

@@ -182,8 +174,7 @@ class StackedFuser(RewriteFuser):

174 `update_attrs` deletes the projections it merges, so any reference to one beyond

175 its call site must be a guard on its existence; folding those to their constant

176 value keeps the rewritten forward off a name that no longer exists. A reference

185 - that is not such a guard raises, so fusion is skipped.

186 - """

177 + that is not such a guard raises, so fusion is skipped."""

178 calls = []

179 for name in names:

180 call, refs = self_call_and_refs(funcdef, name)

vllm/model_executor/models/transformers/fusers/mla.py

+1 -2

@@ -232,8 +232,7 @@ class MLAFuser(StackedFuser):

232 def update_forward(self, module: nn.Module) -> None:

233 """Merge `q_a_proj` and `kv_a_proj` into one fused down proj then split.

234 Bypass the KV expansion method so the compressed latent reaches the `vllm_mla`

235 - attention interface unexpanded.

236 - """

235 + attention interface unexpanded."""

236 funcdef, fn = recover_forward(type(module))

237 if self.has_q_lora:

238 # q_a_proj is usually inside the `else` of `if self.q_lora_rank is None`.

vllm/model_executor/models/transformers/fusers/moe.py

+4 -8

@@ -80,8 +80,7 @@ def _forced_dtype(nodes: Iterable[fx.Node]) -> torch.dtype | None:

80

81 Computations that must run in higher precision say so in their forward (e.g.

82 `hidden_states.type(torch.float32)`), so the dtype is readable from the graph

83 - even when the config does not name it.

84 - """

83 + even when the config does not name it."""

84 for node in nodes:

85 if node.op not in ("call_method", "call_function"):

86 continue

@@ -126,8 +125,7 @@ class SharedExpertMLP(nn.Module):

125 def _moe_block_forward(self: nn.Module, hidden_states: torch.Tensor) -> torch.Tensor:

126 """Standard MoE block forward.

127

129 - Routing and any shared experts are handled inside `self.experts: MoERunner`.

130 - """

128 + Routing and any shared experts are handled inside `self.experts: MoERunner`."""

129 orig_shape = hidden_states.shape

130 hidden_states = hidden_states.reshape(-1, orig_shape[-1])

131 num_tokens = hidden_states.shape[0]

@@ -154,8 +152,7 @@ class MoEBlockFuser:

152 def _match_router(gate: nn.Module) -> tuple[str, torch.dtype | None] | None:

153 """Matches `topk(score(linear(x)))`, `score` being `softmax`/`sigmoid`.

154

157 - Returns the scoring function and the dtype the router computes in.

158 - """

155 + Returns the scoring function and the dtype the router computes in."""

156 state = {name for name, _ in named_state(gate)}

157 if "weight" not in state or state - {"weight", "e_score_correction_bias"}:

158 return None

@@ -283,8 +280,7 @@ class MoEBlockFuser:

280 self, moe_block: nn.Module, prefix: str

281 ) -> SharedExpertMLP | None:

282 """Build the HF shared expert (and its optional gate)

286 - as a `SharedExpertMLP` for vLLM's fused MoE.

287 - """

283 + as a `SharedExpertMLP` for vLLM's fused MoE."""

284 if self.shared_name is None:

285 return None

286 shared_experts = getattr(moe_block, self.shared_name)

vllm/model_executor/models/transformers/fx_utils.py

+11 -22

@@ -43,8 +43,7 @@ def _reference_weight(module: nn.Module) -> torch.Tensor | None:

43 """A weight whose trailing dim is the module's hidden size.

44

45 Linears and 2-D gate weights are `[out, hidden]`; norm weights are

46 - `[hidden]`. Used to fabricate a placeholder input of matching size/dtype.

47 - """

46 + `[hidden]`. Used to fabricate a placeholder input of matching size/dtype."""

47 for child in module.modules():

48 if isinstance(child, nn.Linear):

49 return child.weight

@@ -59,8 +58,7 @@ class _MetaProxy(fx.Proxy):

58

59 Shape questions (`len`, iteration, `.shape` unpacks) are answered by

60 executing each op on the meta values, so PyTorch's meta kernels are the

62 - single source of shape inference — no per-op rules.

63 - """

61 + single source of shape inference — no per-op rules."""

62

63 meta: object = _UNKNOWN

64

@@ -84,8 +82,7 @@ class _MetaAttribute(_MetaProxy, fx.proxy.Attribute):

82 """Attribute proxy (e.g. `x.shape`) carrying its meta value.

83

84 `Proxy.__getattr__` constructs `Attribute` directly, bypassing

87 - `Tracer.proxy`, so the meta value must be grafted on here too.

88 - """

85 + `Tracer.proxy`, so the meta value must be grafted on here too."""

86

87 def __init__(self, root: fx.Proxy, attr: str):

88 super().__init__(root, attr)

@@ -221,8 +218,7 @@ def _leaf_attention_interfaces():

218 """Patch `AttentionInterface.get_interface` so traced forwards see a leaf node.

219

220 `vllm_attention_function` needs runtime context so it is untraceable.

224 - Every interface returns `(attn_output, attn_weights)`.

225 - """

221 + Every interface returns `(attn_output, attn_weights)`."""

222 from transformers.modeling_utils import AttentionInterface

223

224 original = AttentionInterface.get_interface

@@ -300,8 +296,7 @@ def forward_parameters(cls: type[nn.Module]) -> dict[str, inspect.Parameter]:

296 def forward_input_count(cls: type[nn.Module]) -> int:

297 """The number of tensor inputs `cls.forward` declares, excluding `self` and

298 any `*args`/`**kwargs`. Read from the signature, so it is independent of

303 - whether the trace completes (unlike counting placeholders).

304 - """

299 + whether the trace completes (unlike counting placeholders)."""

300 params = list(forward_parameters(cls).values())

301 if not params:

302 return 1 # uninspectable: assume a single input and let matching decide

@@ -378,8 +373,7 @@ def _in_boolean_context(funcdef: ast.FunctionDef, ref: ast.expr) -> bool:

373

374 In these positions the object's identity never escapes, so a reference that

375 is always truthy can be replaced by `True`. `and`/`or` are excluded: they

381 - yield an operand, so the module could escape (`x and self.<name>`).

382 - """

376 + yield an operand, so the module could escape (`x and self.<name>`)."""

377 for node in ast.walk(funcdef):

378 if (

379 isinstance(node, (ast.If, ast.IfExp, ast.While, ast.Assert))

@@ -412,8 +406,7 @@ def bypass_existence_guard(

406 `not self.<name>`): the reference itself to `True`.

407

408 Any other surviving reference escapes the projection's value, which no longer

415 - exists after fusion, so refuse rather than change semantics.

416 - """

409 + exists after fusion, so refuse rather than change semantics."""

410 for node in ast.walk(funcdef):

411 # `self.<name> is (not) None`, either operand order.

412 if not (isinstance(node, ast.Compare) and len(node.ops) == 1):

@@ -474,8 +467,7 @@ def _base_name(node: ast.expr) -> str | None:

467 def _rebound_names(region: list[ast.stmt]) -> set[str]:

468 """Names rebound outright within `region` (`x = ...`, `del x`).

469

477 - These are `Name` nodes in a `Store`/`Del` context.

478 - """

470 + These are `Name` nodes in a `Store`/`Del` context."""

471 return {

472 node.id

473 for stmt in region

@@ -489,8 +481,7 @@ def _inplace_target(node: ast.AST) -> str | None:

481

482 A write through the name (`x[i] = ...`, `x.attr = ...`) or an in-place method

483 call (`x.mul_(...)`) leaves the base name in a `Load` context, so it is not a

492 - plain `Name` store (see `_rebound_names`).

493 - """

484 + plain `Name` store (see `_rebound_names`)."""

485 if isinstance(node, (ast.Attribute, ast.Subscript)) and isinstance(

486 node.ctx, (ast.Store, ast.Del)

487 ):

@@ -711,8 +702,7 @@ def upstream_linear(node: object, module: nn.Module) -> fx.Node | None:

702 Non-linear submodules are transparent too (e.g. the dropout GPT-style

703 attentions apply after their output projection). Never walks through a leaf

704 call (e.g. an attention interface): its inputs are what attention consumes,

714 - not what produced the value.

715 - """

705 + not what produced the value."""

706 stack = [node]

707 seen: set[fx.Node] = set()

708 while stack:

@@ -735,8 +725,7 @@ def downstream_linear(node: fx.Node, module: nn.Module) -> fx.Node | None:

725 """Nearest linear consuming `node`'s output, walking through casts/scalings.

726

727 Never walks through a leaf call (e.g. an attention interface): what crosses

738 - it is consumed by the attention computation, not projected.

739 - """

728 + it is consumed by the attention computation, not projected."""

729 queue = list(node.users)

730 seen: set[fx.Node] = set()

731 while queue:

vllm/model_executor/models/transformers/layers.py

+1 -2

@@ -22,8 +22,7 @@ _VLLM_PKG = "vllm.model_executor.layers"

22

23 def _resolve(module: str, name: str):

24 """Return `name` from the hw-agnostic `module` when enabled and available,

25 - else from vLLM. Logs which source was used.

26 - """

25 + else from vLLM. Logs which source was used."""

26 if envs.VLLM_USE_HW_AGNOSTIC:

27 try:

28 obj = getattr(importlib.import_module(f"{_HW_PKG}.{module}"), name)

vllm/model_executor/models/transformers/moe.py

+2 -4

@@ -75,8 +75,7 @@ class TransformersMoERunner(MoERunner):

75 ) -> torch.Tensor:

76 """In Transformers `experts.forward` will have this signature.

77

78 - We discard any extra kwargs because we cannot use them here.

79 - """

78 + We discard any extra kwargs because we cannot use them here."""

79 # Note: we need to forward through a custom op so the topk_ids

80 # can be transferred without interfering with cudagraphs.

81 return torch.ops.vllm.transformers_moe_forward(

@@ -368,8 +367,7 @@ class MoEMixin(MixtureOfExperts):

367 moe_state: TransformersMoEState,

368 ):

369 """Return `topk_weights` from `gating_output` and the

371 - `topk_ids` we stored in the layer earlier.

372 - """

370 + `topk_ids` we stored in the layer earlier."""

371 topk_weights = gating_output

372 topk_ids = moe_state.topk_ids

373 assert topk_ids is not None

vllm/model_executor/models/transformers/multimodal.py

+8 -16

@@ -424,8 +424,7 @@ class LegacyMultiModalProcessor(_MultiModalProcessorBase):

424 out_mm_kwargs: MultiModalKwargsItems,

425 ) -> Sequence[PromptUpdate]:

426 """Empty, because `apply` writes the placeholder ranges itself rather than

427 - deriving them from updates.

428 - """

427 + deriving them from updates."""

428 return []

429

430 def _get_mm_token_ids(self, modality: str) -> list[int]:

@@ -459,8 +458,7 @@ class LegacyMultiModalProcessor(_MultiModalProcessorBase):

458 num_audios: int,

459 ) -> dict[str, list[PlaceholderRange]]:

460 """Take each contiguous run of the audio token as one item's placeholder,

462 - and record how many tokens the run holds.

463 - """

461 + and record how many tokens the run holds."""

462 audio_token_ids = self._get_mm_token_ids("audio")

463 prompt_tensor = torch.tensor(prompt_ids)

464 is_audio = torch.isin(prompt_tensor, torch.tensor(audio_token_ids))

@@ -518,8 +516,7 @@ class LegacyMultiModalProcessor(_MultiModalProcessorBase):

516 mm_token_type_ids: torch.Tensor | None,

517 ) -> dict[str, list[PlaceholderRange]]:

518 """Split the positions the processor marks as image into one placeholder per

521 - item, sized by the token count it reports for each image.

522 - """

519 + item, sized by the token count it reports for each image."""

520 hf_processor = self.info.get_hf_processor(**hf_processor_mm_kwargs)

521 if mm_token_type_ids is None:

522 raise ValueError(

@@ -572,8 +569,7 @@ class LegacyMultiModalProcessor(_MultiModalProcessorBase):

569 timing_ctx: TimingContext,

570 ) -> MultiModalInput:

571 """Process the prompt and every multi-modal item in one HF processor call,

575 - then read the placeholder ranges out of the token ids it returns.

576 - """

572 + then read the placeholder ranges out of the token ids it returns."""

573 prompt = inputs.prompt

574 mm_items = inputs.mm_data_items

575 hf_processor_mm_kwargs = inputs.hf_processor_mm_kwargs

@@ -711,8 +707,7 @@ class OffsetsMultiModalProcessor(_MultiModalProcessorBase):

707 out_mm_kwargs: MultiModalKwargsItems,

708 ) -> Sequence[PromptUpdate]:

709 """Replace each modality's placeholder token with the token ids that item's

714 - replacement text encodes to, marking which of them hold embeddings.

715 - """

710 + replacement text encodes to, marking which of them hold embeddings."""

711 hf_processor = self.info.get_hf_processor(**hf_processor_mm_kwargs)

712 tokenizer = self.info.get_tokenizer()

713

@@ -907,8 +902,7 @@ class MultiModalMixin(SupportsMultiModal, SupportsMRoPE):

902 self, model: "PreTrainedModel"

903 ) -> dict[str, type["PreTrainedModel"]]:

904 """Modalities whose encoder cannot be told apart from the model itself are

910 - omitted, as are those `get_encoder` rejects.

911 - """

905 + omitted, as are those `get_encoder` rejects."""

906 encoder_classes: dict[str, type[PreTrainedModel]] = {}

907 for modality in _MODALITY_TO_TOKEN_TYPE_ID:

908 try:

@@ -1027,8 +1021,7 @@ class MultiModalMixin(SupportsMultiModal, SupportsMRoPE):

1021 def get_language_model(self) -> torch.nn.Module:

1022 """Transformers modeling backend multimodal classes do not contain a separate

1023 vLLM language model class. Therefore, in order to return a language model vLLM

1030 - class, we use a wrapper to give `self` the same interface as a text model.

1031 - """

1024 + class, we use a wrapper to give `self` the same interface as a text model."""

1025 # Exclude self and object

1026 bases = self.__class__.mro()[1:-1]

1027 # Keep only classes defined in `vllm.model_executor.models.transformers`

@@ -1182,8 +1175,7 @@ class MultiModalMixin(SupportsMultiModal, SupportsMRoPE):

1175 ) -> dict[str, Any]:

1176 """Narrow the entries of `kwargs` that hold one row per item down to the item

1177 at `index`. Length is all there is to match on, so an unrelated entry of the

1185 - same length is narrowed too.

1186 - """

1178 + same length is narrowed too."""

1179 return {

1180 key: value[index : index + 1]

1181 if isinstance(value, (torch.Tensor, list)) and len(value) == num_items

vllm/model_executor/models/transformers/utils.py

+2 -4

@@ -201,8 +201,7 @@ def attrsetter(attr: str) -> Callable[[object, object], None]:

201

202 class _UninitializedEmbedding(nn.Embedding):

203 """Make `__init__` inert, so that `VocabParallelEmbedding.__init__`'

204 - call to `super().__init__` does not invoke `nn.Embedding.__init__`.

205 - """

204 + call to `super().__init__` does not invoke `nn.Embedding.__init__`."""

205

206 def __init__(self):

207 pass

@@ -210,8 +209,7 @@ class _UninitializedEmbedding(nn.Embedding):

209

210 class _VocabParallelEmbeddingBase(VocabParallelEmbedding, _UninitializedEmbedding):

211 """Orders `VocabParallelEmbedding` ahead of `nn.Embedding` in the MRO, so that

213 - `super().forward(...)` in an `nn.Embedding` subclass reaches vLLM's embedding.

214 - """

212 + `super().forward(...)` in an `nn.Embedding` subclass reaches vLLM's embedding."""

213

214

215 @lru_cache

vllm/model_executor/models/utils.py

+2 -4

@@ -47,8 +47,7 @@ ShardId: TypeAlias = str | int | tuple[int, ...]

47 class WeightsMapper:

48 """Maps the name of each weight if they match the following patterns.

49

50 - If a key maps to a value of `None`, the corresponding weight is ignored.

51 - """

50 + If a key maps to a value of `None`, the corresponding weight is ignored."""

51

52 orig_to_new_renaming: list["WeightRenaming"] = field(default_factory=list)

53 orig_to_new_regex: Mapping[re.Pattern, str | None] = field(default_factory=dict)

@@ -172,8 +171,7 @@ class WeightsMapper:

171 rather than being rewritten to the stacked vLLM name (`qkv_proj`). Mappings to

172 `None` are dropped because "do not load this weight" is meaningless to such a

173 consumer, and applying it would silently shrink a quantization config's ignore

175 - list or make LoRA name parsing fail.

176 - """

174 + list or make LoRA name parsing fail."""

175 remove_none = lambda d: {k: v for k, v in d.items() if v is not None}

176 return replace(

177 self,

vllm/model_executor/models/vision.py

+1 -2

@@ -116,8 +116,7 @@ def get_multimodal_config() -> MultiModalConfig | None:

116 """Return the current ``MultiModalConfig``, or ``None`` when no engine

117 config context is active (e.g., during unit tests) or when the current

118 ``model_config`` does not carry a ``multimodal_config`` (e.g., minimal

119 - stubs used in tests).

120 - """

119 + stubs used in tests)."""

120 vllm_config = get_current_vllm_config_or_none()

121 if vllm_config is None or vllm_config.model_config is None:

122 return None

vllm/model_executor/utils.py

+1 -2

@@ -174,8 +174,7 @@ def get_moe_expert_mapping(

174

175 It will be retrieved from the first module that has a `get_expert_mapping` method.

176 If the model manually implements `get_expert_mapping`, it will be used.

177 - Otherwise, it will use the first RoutedExperts layer.

178 - """

177 + Otherwise, it will use the first RoutedExperts layer."""

178 for _, module in model.named_modules():

179 get_mapping = getattr(module, "get_expert_mapping", None)

180 if get_mapping is not None:

vllm/model_executor/warmup/qwen4_exp_qsa_warmup.py

+1 -2

@@ -18,8 +18,7 @@ logger = init_logger(__name__)

18

19 def qwen4_exp_qsa_triton_warmup(worker: "Worker") -> None:

20 """Warm every reachable QSA specialization: indexer decode-query-length

21 - profiles plus the sparse attention split-K/merge configs.

22 - """

21 + profiles plus the sparse attention split-K/merge configs."""

22 qsa_module = sys.modules.get("vllm.models.qwen4_exp.nvidia.indexer_qsa")

23 attn_module = sys.modules.get("vllm.models.qwen4_exp.nvidia.qsa")

24 if qsa_module is None or attn_module is None:

vllm/models/deepseek_v32/common/kernels.py

+1 -2

@@ -1059,8 +1059,7 @@ def _fused_eh_norm_kernel(

1059 ):

1060 """MTP input fusion: zero embeds at position 0, RMSNorm(embeds) with enorm

1061 and RMSNorm(prev_hidden) with hnorm, written side-by-side into ``out``

1062 - ([N, 2H]) ready for the eh_proj GEMM. Replaces where + 2x RMSNorm + cat.

1063 - """

1062 + ([N, 2H]) ready for the eh_proj GEMM. Replaces where + 2x RMSNorm + cat."""

1063 tok = tl.program_id(0)

1064 off = tl.arange(0, BLOCK)

1065 mask = off < H

vllm/models/deepseek_v4/amd/model.py

+1 -2

@@ -1425,8 +1425,7 @@ class DeepseekV4ForCausalLM(nn.Module, SupportsPP, SupportsEagle3):

1425 def get_mtp_target_hidden_states(self) -> torch.Tensor | None:

1426 """Pre-hc_head residual stream buffer (max_num_batched_tokens,

1427 hc_mult * hidden_size) for the MTP draft model. Populated by

1428 - forward(); valid after each target step.

1429 - """

1428 + forward(); valid after each target step."""

1429 return getattr(self.model, "_mtp_hidden_buffer", None)

1430

1431 def load_weights(self, weights: Iterable[tuple[str, torch.Tensor]]) -> set[str]:

vllm/models/deepseek_v4/common/mm_preprocess.py

+3 -6

@@ -91,8 +91,7 @@ def validate_image_sentinel_ids(tokenizer) -> None:

91

92 def grid_tokens(best_height, best_width, patch_size, downsample_ratio):

93 """Number of LLM tokens the aligner grid occupies (N-layout, including

94 - row/align padding).

95 - """

94 + row/align padding)."""

95 n_llm_h = math.ceil((best_height // patch_size) / downsample_ratio)

96 n_llm_w = math.ceil((best_width // patch_size) / downsample_ratio)

97 num_tokens = n_llm_h * (n_llm_w + 1) + 2

@@ -197,8 +196,7 @@ def load_image(

196

197 def build_image_block(n_llm_h: int, n_llm_w: int, start_pos: int):

198 """Builds the N-layout token types (final order) and the aligner-row order

200 - for IMAGE slots.

201 - """

199 + for IMAGE slots."""

200 compress_pad = COMPRESS_PAD_TO - 1 - start_pos % COMPRESS_PAD_TO

201 pad_h = n_llm_h % 2

202 rows = n_llm_h + pad_h

@@ -241,8 +239,7 @@ def build_image_block_pad_free(n_llm_h: int, n_llm_w: int):

239

240 class DeepseekV4VLImageProcessor:

241 """Per-image transform (the PIL-input equivalent of the reference

244 - ``load_image``).

245 - """

242 + ``load_image``)."""

243

244 def __init__(self, config: DeepseekV4Config) -> None:

245 super().__init__()

vllm/models/deepseek_v4/common/ops/fused_indexer_q.py

+1 -2

@@ -312,8 +312,7 @@ class FusedIndexerQRopeQuantTritonKernel(

312 def _indexer_weights_out_dtypes(vllm_config: Any) -> tuple[torch.dtype, ...]:

313 """Weights dtypes the model's indexer layers ask for: fp32 for the dense

314 scoring kernels, plus bf16 when the DeepSeek V4.1 sparse-logits indexer

315 - (`SparseMQAIndexer`) is enabled.

316 - """

315 + (`SparseMQAIndexer`) is enabled."""

316 if vllm_config.attention_config.indexer_sparse_logits:

317 return (torch.float32, torch.bfloat16)

318 return (torch.float32,)

vllm/models/deepseek_v4/cpu/cpu_compressor.py

+1 -2

@@ -42,8 +42,7 @@ class DeepseekV4CPUCompressor(DeepseekCompressor):

42 weight loading (see

43 ``DeepseekV4CPUAttention.process_weights_after_loading``) -- the CPU

44 kernel requires fp32/contiguous but the checkpoint loads this weight

45 - in bf16, and it never changes after loading.

46 - """

45 + in bf16, and it never changes after loading."""

46 self._norm_weight_fp32 = self.norm.weight.to(torch.float32).contiguous()

47

48 def forward(

vllm/models/deepseek_v4/cpu/cpu_mla.py

+1 -2

@@ -118,8 +118,7 @@ class DeepseekV4CPUSparseSWAMetadataBuilder(DeepseekSparseSWAMetadataBuilder):

118 ) -> dict[str, FlashMLASchedMeta | None]:

119 """CPU never runs the FlashMLA tile-scheduler planner (that's a CUDA

120 C++ decode-path concern) -- always return the all-``None`` sentinel

121 - the shared base's own CPU branch returns.

122 - """

121 + the shared base's own CPU branch returns."""

122 return {

123 _LAYER_TYPE_SWAONLY: None,

124 _LAYER_TYPE_C4A: None,

vllm/models/deepseek_v4/cpu/cpu_sparse.py

+8 -16

@@ -176,15 +176,13 @@ def _fused_indexer_q_rope_quant_cpu(

176

177 class _NoOpEvent:

178 """Stand-in for ``torch.cuda.Event`` on CPU: since CPU never sets an aux

179 - stream list, these events are constructed but never actually used.

180 - """

179 + stream list, these events are constructed but never actually used."""

180

181

182 class DeepseekV4CPUIndexerCache(DeepseekV4IndexerCache):

183 """CPU indexer K-cache descriptor: same fields as the shared base, just

184 pointing ``get_attn_backend()`` at ``DeepseekV4CPUIndexerBackend``

186 - instead of the shared, CUDA/XPU-oriented ``DeepseekV4IndexerBackend``.

187 - """

185 + instead of the shared, CUDA/XPU-oriented ``DeepseekV4IndexerBackend``."""

186

187 def get_attn_backend(self) -> type[AttentionBackend]:

188 return DeepseekV4CPUIndexerBackend

@@ -211,8 +209,7 @@ class DeepseekV4CPUIndexer(DeepseekV4Indexer):

209 compressor run straight-line instead of through

210 ``maybe_execute_in_parallel``; the short-context check drops the

211 CUDA-only ``is_current_stream_capturing()`` guard (always False

214 - here).

215 - """

212 + here)."""

213 compressor = self.compressor

214

215 attn_metadata = get_forward_context().attn_metadata

@@ -312,8 +309,7 @@ class DeepseekV4CPUAttention(DeepseekV4Attention):

309 """Snapshot and pack ``wo_a``'s weight into a bf16 copy for

310 ``bmm_cpu`` before the FP8 kernel's own

311 ``process_weights_after_loading`` VNNI-repacks it in place for

315 - row-major reads ``_o_proj`` never does.

316 - """

312 + row-major reads ``_o_proj`` never does."""

313 wo_a = self.wo_a

314 quant_method = wo_a.quant_method

315 orig_pwal = quant_method.process_weights_after_loading

@@ -404,8 +400,7 @@ class DeepseekV4CPUAttention(DeepseekV4Attention):

400 ) -> torch.Tensor:

401 """CPU override: only the fp8_ds_mla (uint8) SWA cache layout is

402 supported here, so the base method's bf16/per-tensor-fp8 branches

407 - are dropped.

408 - """

403 + are dropped."""

404 if not isinstance(attn_metadata, dict):

405 # Profile run: kernel doesn't fire; produce a padded tensor so

406 # downstream FlashMLA gets the right shape.

@@ -450,8 +445,7 @@ class DeepseekV4CPUAttention(DeepseekV4Attention):

445 ) -> tuple[torch.Tensor, torch.Tensor | None, torch.Tensor]:

446 """CPU override: two ``RMSNorm`` calls instead of the shared

447 ``fused_q_kv_rmsnorm``, whose raw ``@triton.jit`` kernel fails to

453 - link under triton-cpu (``undefined symbol: __truncdfbf2``).

454 - """

448 + link under triton-cpu (``undefined symbol: __truncdfbf2``)."""

449 qr, kv = qr_kv.split([self.q_lora_rank, self.head_dim], dim=-1)

450 return self.q_norm(qr), None, self.kv_norm(kv)

451

@@ -469,8 +463,7 @@ class DeepseekV4CPUAttention(DeepseekV4Attention):

463 ) -> None:

464 """CPU override: no aux streams, so query-projection+KV-insert, the

465 indexer, and the compressor run straight-line instead of through

472 - execute_in_parallel/maybe_execute_in_parallel.

473 - """

466 + execute_in_parallel/maybe_execute_in_parallel."""

467 attn_metadata = get_forward_context().attn_metadata

468 indexer = self.indexer

469 compressor = self.compressor

@@ -519,8 +512,7 @@ class DeepseekV4CPUAttention(DeepseekV4Attention):

512 out: torch.Tensor,

513 ) -> None:

514 """CPU override: identical body, minus ``@eager_break_during_capture``

522 - (this platform never captures a CUDA graph, so it was a no-op).

523 - """

515 + (this platform never captures a CUDA graph, so it was a no-op)."""

516 if self.indexer is not None and index_q is not None:

517 assert index_weights is not None

518 # index_q_scale is always None on CPU (FP8-only path).

vllm/models/deepseek_v4/cpu/model.py

+1 -2

@@ -907,8 +907,7 @@ class DeepseekV4ForCausalLM(nn.Module, SupportsPP, SupportsEagle3):

907 def get_mtp_target_hidden_states(self) -> torch.Tensor | None:

908 """Pre-hc_head residual stream buffer (max_num_batched_tokens,

909 hc_mult * hidden_size) for the MTP draft model. Populated by

910 - forward(); valid after each target step.

911 - """

910 + forward(); valid after each target step."""

911 return getattr(self.model, "_mtp_hidden_buffer", None)

912

913 def load_weights(self, weights: Iterable[tuple[str, torch.Tensor]]) -> set[str]:

vllm/models/deepseek_v4/nvidia/fi_moe.py

+1 -2

@@ -159,8 +159,7 @@ class DeepseekV4MegaMoEExpertsFI(DeepseekV4MegaMoEExperts):

159 def _realloc_nvfp4_params(self) -> None:

160 """Swap the mx-recipe scale params for the NVFP4 checkpoint's:

161 fp8-e4m3 per-16 block scales plus the per-tensor second-level

162 - scales (weight_scale_2, input_scale) modelopt exports.

163 - """

162 + scales (weight_scale_2, input_scale) modelopt exports."""

163 n_e = self.num_local_experts

164 inter = self.intermediate_size

165 hidden = self.hidden_size

vllm/models/deepseek_v4/nvidia/flashinfer_sparse.py

+1 -2

@@ -56,8 +56,7 @@ def _get_flashinfer_dsv4_workspace(device: torch.device) -> torch.Tensor:

56 def _packed_block_span(pool: torch.Tensor) -> int:

57 """Per-block stride of ``pool`` in tokens (``stride(0)//stride(-2)``): ==

58 block_size for unpacked KV, larger when packed (#44577). Raises if not

59 - token-aligned.

60 - """

59 + token-aligned."""

60 block_stride = pool.stride(0)

61 token_stride = pool.stride(-2)

62 if block_stride % token_stride != 0:

vllm/models/deepseek_v4/nvidia/model.py

+1 -2

@@ -1985,8 +1985,7 @@ class DeepseekV4ForCausalLM(

1985 def get_mtp_target_hidden_states(self) -> torch.Tensor | None:

1986 """Pre-hc_head residual stream buffer (max_num_batched_tokens,

1987 hc_mult * hidden_size) for the MTP draft model. Populated by

1988 - forward(); valid after each target step.

1989 - """

1988 + forward(); valid after each target step."""

1989 return getattr(self.model, "_mtp_hidden_buffer", None)

1990

1991 def load_weights(self, weights: Iterable[tuple[str, torch.Tensor]]) -> set[str]:

vllm/models/deepseek_v4/xpu/model.py

+1 -2

@@ -1378,8 +1378,7 @@ class DeepseekV4ForCausalLM(nn.Module, SupportsPP, SupportsEagle3, SupportsLoRA)

1378 def get_mtp_target_hidden_states(self) -> torch.Tensor | None:

1379 """Pre-hc_head residual stream buffer (max_num_batched_tokens,

1380 hc_mult * hidden_size) for the MTP draft model. Populated by

1381 - forward(); valid after each target step.

1382 - """

1381 + forward(); valid after each target step."""

1382 return getattr(self.model, "_mtp_hidden_buffer", None)

1383

1384 def load_weights(self, weights: Iterable[tuple[str, torch.Tensor]]) -> set[str]:

vllm/models/deepseek_v41/amd/model.py

+2 -4

@@ -1062,8 +1062,7 @@ class DeepseekV41LLMForCausalLM(

1062 @property

1063 def token_lookback_depth(self) -> int:

1064 """Tokens before a chunk start the engram hash needs; the model runner

1065 - passes them as `lookback_token_ids`.

1066 - """

1065 + passes them as `lookback_token_ids`."""

1066 engram_hash = self.model.engram_hash

1067 return engram_hash.lookback_depth if engram_hash is not None else 0

1068

@@ -1087,8 +1086,7 @@ class DeepseekV41LLMForCausalLM(

1086 def get_mtp_target_hidden_states(self) -> torch.Tensor | None:

1087 """Pre-collapse residual stream buffer (max_num_batched_tokens,

1088 hc_mult * hidden_size) for the MTP draft model. Populated by

1090 - forward(); valid after each target step.

1091 - """

1089 + forward(); valid after each target step."""

1090 return getattr(self.model, "_mtp_hidden_buffer", None)

1091

1092 def load_weights(self, weights: Iterable[tuple[str, torch.Tensor]]) -> set[str]:

vllm/models/deepseek_v41/amd/vl_model.py

+1 -2

@@ -209,8 +209,7 @@ class DeepseekV41ForCausalLM(nn.Module, SupportsMultiModal, SupportsPP, Supports

209 self, image_embeds: torch.Tensor, types: torch.Tensor

210 ) -> torch.Tensor:

211 """Full image span: aligner rows at IMAGE slots, the learned

212 - delimiter vectors at IMAGE_START/IMAGE_NEW_LINE/IMAGE_END.

213 - """

212 + delimiter vectors at IMAGE_START/IMAGE_NEW_LINE/IMAGE_END."""

213 types = types.to(image_embeds.device)

214 span = image_embeds.new_empty(types.numel(), image_embeds.shape[-1])

215 dtype = image_embeds.dtype

vllm/models/deepseek_v41/attention.py

+1 -2

@@ -1066,8 +1066,7 @@ class DeepseekV4Attention(nn.Module, AttentionLayerBase, ABC):

1066

1067 def _compressed_kv_cache(self) -> torch.Tensor:

1068 """The compressed-KV cache tensor of this layer's kv source (own

1069 - cache for kv-source layers).

1070 - """

1069 + cache for kv-source layers)."""

1070 if self.is_kv_source:

1071 return self.kv_cache

1072 assert self.compressed_cache_prefix is not None

vllm/models/deepseek_v41/common/engram.py

+3 -6

@@ -564,8 +564,7 @@ def _engram_head_shard_weight_loader(

564 param: torch.nn.Parameter, loaded_weight: torch.Tensor

565 ) -> None:

566 """Load this rank's complete head buckets. ue8m0 scales arrive as

567 - float8_e8m0fnu; keep the raw bytes (the param stores uint8).

568 - """

567 + float8_e8m0fnu; keep the raw bytes (the param stores uint8)."""

568 part_rows = param.shape[0]

569 if loaded_weight.dtype == torch.float8_e8m0fnu:

570 loaded_weight = loaded_weight.view(torch.uint8)

@@ -740,8 +739,7 @@ class ParallelEngramEmbedding(nn.Module):

739

740 def forward(self, indices: torch.Tensor) -> torch.Tensor:

741 """indices: [num_tokens, n_hash_cols] -> [num_tokens, n_hash_cols, dim]

743 - bf16, gathered from all shards for this replica's tokens.

744 - """

742 + bf16, gathered from all shards for this replica's tokens."""

743 out = torch.empty(

744 (indices.shape[0], self.part_n_hash_cols, self.dim),

745 dtype=torch.bfloat16,

@@ -1002,8 +1000,7 @@ class Engram(nn.Module):

1000 ) -> torch.Tensor:

1001 """hidden_states: [T, hc_mult, dim]; hash_ids: [T, n_hash_cols] (all

1002 tokens, pre sequence-parallel shard); token_mask: [T], False shuts

1005 - the gate so those positions pass through untouched.

1006 - """

1003 + the gate so those positions pass through untouched."""

1004 kv = self.wkv(self.embed(hash_ids).flatten(-2))

1005 num_kv_tokens = hash_ids.shape[0]

1006 assert token_mask is None or token_mask.shape == (num_kv_tokens,)

vllm/models/deepseek_v41/common/mm_preprocess.py

+3 -6

@@ -82,8 +82,7 @@ def num_image_tokens(n_llm_h: int, n_llm_w: int) -> int:

82

83 def solve_resize_ratio(height, width, patch_size, downsample_ratio, max_n_token):

84 """Largest aspect-preserving pixel size whose token grid still fits in

85 - max_n_token. Returns (best_height, best_width).

86 - """

85 + max_n_token. Returns (best_height, best_width)."""

86 r = height / width

87 max_w_float = math.sqrt((max_n_token - 2) / r + 0.25) - 0.5

88 max_h_float = max_w_float * r

@@ -106,8 +105,7 @@ def safe_resize(

105 height, width, best_height, best_width, patch_size, downsample_ratio, max_n_token

106 ):

107 """Shrink the pixel size until the image costs at most max_n_token LLM

109 - tokens.

110 - """

108 + tokens."""

109 n_llm_h, n_llm_w = llm_grid(best_height, best_width, patch_size, downsample_ratio)

110 if num_image_tokens(n_llm_h, n_llm_w) > max_n_token:

111 best_height, best_width = solve_resize_ratio(

@@ -173,8 +171,7 @@ def image_token_types(n_llm_h: int, n_llm_w: int) -> torch.Tensor:

171

172 class DeepseekV4VLImageProcessor:

173 """Per-image transform (the PIL-input equivalent of the reference

176 - ``load_image``).

177 - """

174 + ``load_image``)."""

175

176 def __init__(self, config: DeepseekV41Config) -> None:

177 super().__init__()

vllm/models/deepseek_v41/nvidia/flashinfer_sparse.py

+1 -2

@@ -54,8 +54,7 @@ def _get_flashinfer_dsv4_workspace(device: torch.device) -> torch.Tensor:

54 def _packed_block_span(pool: torch.Tensor) -> int:

55 """Per-block stride of ``pool`` in tokens (``stride(0)//stride(-2)``): ==

56 block_size for unpacked KV, larger when packed (#44577). Raises if not

57 - token-aligned.

58 - """

57 + token-aligned."""

58 block_stride = pool.stride(0)

59 token_stride = pool.stride(-2)

60 if block_stride % token_stride != 0:

vllm/models/deepseek_v41/nvidia/model.py

+2 -4

@@ -1179,8 +1179,7 @@ class DeepseekV41LLMForCausalLM(

1179 @property

1180 def token_lookback_depth(self) -> int:

1181 """Tokens before a chunk start the engram hash needs; the model runner

1182 - passes them as `lookback_token_ids`.

1183 - """

1182 + passes them as `lookback_token_ids`."""

1183 engram_hash = self.model.engram_hash

1184 return engram_hash.lookback_depth if engram_hash is not None else 0

1185

@@ -1204,8 +1203,7 @@ class DeepseekV41LLMForCausalLM(

1203 def get_mtp_target_hidden_states(self) -> torch.Tensor | None:

1204 """Pre-collapse residual stream buffer (max_num_batched_tokens,

1205 hc_mult * hidden_size) for the MTP draft model. Populated by

1207 - forward(); valid after each target step.

1208 - """

1206 + forward(); valid after each target step."""

1207 return getattr(self.model, "_mtp_hidden_buffer", None)

1208

1209 def load_weights(self, weights: Iterable[tuple[str, torch.Tensor]]) -> set[str]:

vllm/models/deepseek_v41/nvidia/vl_model.py

+1 -2

@@ -209,8 +209,7 @@ class DeepseekV41ForCausalLM(nn.Module, SupportsMultiModal, SupportsPP, Supports

209 self, image_embeds: torch.Tensor, types: torch.Tensor

210 ) -> torch.Tensor:

211 """Full image span: aligner rows at IMAGE slots, the learned

212 - delimiter vectors at IMAGE_START/IMAGE_NEW_LINE/IMAGE_END.

213 - """

212 + delimiter vectors at IMAGE_START/IMAGE_NEW_LINE/IMAGE_END."""

213 types = types.to(image_embeds.device)

214 span = image_embeds.new_empty(types.numel(), image_embeds.shape[-1])

215 dtype = image_embeds.dtype

vllm/models/glm5next/amd/ops/kpool_compress.py

+1 -2

@@ -98,8 +98,7 @@ def _kpool_softmax_rotate_write_cache_kernel(

98 BLOCK_D: tl.constexpr,

99 ):

100 """One program per pool. softmax(slot_score+ape)-weighted sum of slot_k ->

101 - Hadamard-128 -> per-vector fp8 absmax quant -> write to cache at ``loc``.

102 - """

101 + Hadamard-128 -> per-vector fp8 absmax quant -> write to cache at ``loc``."""

102 row = tl.program_id(0)

103 do_write = True

104 if HAS_WRITE_MASK:

vllm/models/glm5next/common/kda.py

+2 -4

@@ -130,8 +130,7 @@ def _resolve_kda_prefill_backend(

130 """Pick the chunked-prefill kernel: FlashKDA (fused CUDA, ~2-4x faster on

131 SM90/SM10x/SM12x for bf16, head_dim 128 and a bounded gate) or the Triton

132 ``chunk_kda_with_fused_gate`` path. ``backend`` comes from

133 - ``additional_config.kda_prefill_backend`` (auto / triton / flashkda).

134 - """

133 + ``additional_config.kda_prefill_backend`` (auto / triton / flashkda)."""

134 if backend not in ("auto", "triton", "flashkda"):

135 raise ValueError(f"Unsupported KDA prefill backend: {backend}")

136 capability = current_platform.get_device_capability()

@@ -369,8 +368,7 @@ class Glm5NextLinearAttention(GatedDeltaNetAttention):

368 bounded gate ``lower_bound * sigmoid(exp(A_log) * (g + dt_bias))``,

369 matching ``chunk_kda_with_fused_gate(..., safe_gate=True)``. Writes the

370 attention output into ``out`` (a workspace buffer when ``None``) and

372 - returns ``(out, final_state)``.

373 - """

371 + returns ``(out, final_state)``."""

372 assert self._flashkda_buffer_specs is not None

373 final_state, workspace, workspace_out = (

374 current_workspace_manager().get_simultaneous(*self._flashkda_buffer_specs)

vllm/models/glm5next/common/multimodal.py

+1 -2

@@ -720,8 +720,7 @@ class Glm5NextMultiModalProcessor(Glm4vMultiModalProcessor):

720 (image token repeat, video frame/timestamp structure) is owned by vLLM's

721 prompt-update machinery — the inherited ``_get_prompt_updates`` builds

722 the replacement content and the placeholder scan validates against

723 - exactly that.

724 - """

723 + exactly that."""

724

725 def _hf_processor_applies_updates(

726 self,

vllm/models/glm5next/common/sparse_indexer.py

+2 -4

@@ -138,8 +138,7 @@ def _gather_workspace_shapes(

138 """Return ((values_shape, values_dtype), (scales_shape, scales_dtype)) for

139 the K-gather workspace. FP8 path: (T, head_dim) fp8 + (T, 4) uint8 fp32

140 scales. MXFP4 path: (T, head_dim // 2) uint8 packed mxfp4 +

141 - (T, head_dim // MXFP4_BLOCK_SIZE) uint8 ue8m0 scales.

142 - """

141 + (T, head_dim // MXFP4_BLOCK_SIZE) uint8 ue8m0 scales."""

142 if use_fp4_cache:

143 return (

144 ((total_seq_lens, head_dim // 2), torch.uint8),

@@ -157,8 +156,7 @@ def kv_cache_as_quant_view(

156 use_fp4_cache: bool,

157 ) -> torch.Tensor:

158 """4D ``[num_blocks, block_size, 1, head_width]`` view expected by

160 - DeepGEMM, from the 3D indexer kv-cache allocation.

161 - """

159 + DeepGEMM, from the 3D indexer kv-cache allocation."""

160 if use_fp4_cache:

161 assert kv_cache.ndim == 3 and kv_cache.dtype == torch.uint8

162 num_blocks, block_size, _ = kv_cache.shape

vllm/models/glm5next/nvidia/ops/fused_eh_norm.py

+1 -2

@@ -31,8 +31,7 @@ def _fused_eh_norm_kernel(

31 ):

32 """MTP input fusion: zero embeds at position 0, RMSNorm(embeds) with enorm

33 and RMSNorm(prev_hidden) with hnorm, written side-by-side into ``out``

34 - ([N, 2H]) ready for the eh_proj GEMM. Replaces where + 2x RMSNorm + cat.

35 - """

34 + ([N, 2H]) ready for the eh_proj GEMM. Replaces where + 2x RMSNorm + cat."""

35 tok = tl.program_id(0)

36 off = tl.arange(0, BLOCK)

37 mask = off < H

vllm/models/glm5next/nvidia/ops/kpool_compress.py

+1 -2

@@ -163,8 +163,7 @@ def _kpool_softmax_rotate_write_cache_kernel(

163 BLOCK_D: tl.constexpr,

164 ):

165 """One program per pool. softmax(slot_score+ape)-weighted sum of slot_k ->

166 - Hadamard-128 -> per-vector fp8 absmax quant -> write to cache at ``loc``.

167 - """

166 + Hadamard-128 -> per-vector fp8 absmax quant -> write to cache at ``loc``."""

167 row = tl.program_id(0)

168 do_write = True

169 if HAS_WRITE_MASK:

vllm/models/inkling/amd/model.py

+1 -2

@@ -72,8 +72,7 @@ def _sconv_add_norm(

72

73 ROCm uses the portable collective path. The Lamport P2P implementation in

74 the NVIDIA model relies on CUDA GDC and CUDA-specific symmetric-memory

75 - publication semantics which cannot be linked into a gfx950 HSACO.

76 - """

75 + publication semantics which cannot be linked into a gfx950 HSACO."""

76 norm_w = norm.weight if norm is not None else None

77 eps = norm.variance_epsilon if norm is not None else 0.0

78

vllm/models/inkling/amd/moe.py

+3 -6

@@ -192,8 +192,7 @@ def inkling_gate_select(

192

193 class InklingGate(nn.Module):

194 """Sigmoid gate with selection bias, log-sigmoid renorm after top-k, and

195 - global scale (the served checkpoint's only configuration).

196 - """

195 + global scale (the served checkpoint's only configuration)."""

196

197 def __init__(

198 self,

@@ -247,8 +246,7 @@ class InklingGate(nn.Module):

246 self, gating_output: torch.Tensor

247 ) -> tuple[torch.Tensor, torch.Tensor]:

248 """Full selection: (weights, ids) of [T, K + S]. The first K entries

250 - are the routed top-k; the S trailing entries are the sink gammas.

251 - """

249 + are the routed top-k; the S trailing entries are the sink gammas."""

250 return inkling_gate_select(

251 gating_output,

252 self.n_total_experts,

@@ -269,8 +267,7 @@ class InklingGate(nn.Module):

267 def _inkling_moe_ep_size() -> int:

268 """EP size the MoERunner layer will run with (mirrors

269 FusedMoEParallelConfig.make: experts shard over tp * dp * pcp when

272 - expert parallelism is enabled).

273 - """

270 + expert parallelism is enabled)."""

271 parallel_config = get_current_vllm_config().parallel_config

272 if not parallel_config.enable_expert_parallel:

273 return 1

vllm/models/inkling/amd/mtp.py

+3 -6

@@ -166,8 +166,7 @@ class InklingMultiTokenPredictor(nn.Module):

166 ) -> torch.Tensor:

167 """Draft-prefill embedding: fused gather + backbone embed_norm, then

168 the target's tower embeddings scattered in unnormed (the backbone

169 - convention — MM embeds are merged after embed_norm).

170 - """

169 + convention — MM embeds are merged after embed_norm)."""

170 norm = self.backbone_embed_norm

171 embeds = embed_rmsnorm(

172 input_ids,

@@ -196,8 +195,7 @@ class InklingMultiTokenPredictor(nn.Module):

195 """The depth layer's [rmsnorm(hidden) | embed_norm(embed)] input in one

196 launch: embedding row gather + the backbone embed_norm + the depth

197 embed_norm chain on one side, hidden_norm on the other, written

199 - straight into the cat buffer.

200 - """

198 + straight into the cat buffer."""

199 hidden_w = layer.hidden_norm.weight

200 embed_w = layer.embed_norm.weight

201 eps = layer.hidden_norm.variance_epsilon

@@ -329,8 +327,7 @@ class InklingMTP(nn.Module, SupportsMultiModalEmbeddings):

327 """Greedy draft tokens via rank-local argmax + tiny (value, index)

328 reduction — no full-vocab logits all-gather. The muP divisor is a

329 positive scalar, so the argmax is invariant and the scaling is

332 - skipped entirely.

333 - """

330 + skipped entirely."""

331 return self.logits_processor.get_top_tokens(self.lm_head, hidden_states)

332

333 def load_weights(self, weights: Iterable[tuple[str, torch.Tensor]]) -> set[str]:

vllm/models/inkling/amd/ops/norm.py

+2 -4

@@ -238,8 +238,7 @@ def embed_rmsnorm(

238 embedding-table gather.

239 ``chain_weight`` additionally emits ``rmsnorm(out, chain_weight)`` (the

240 first decoder layer's pre-attention norm) as a second output, still one

241 - launch. Bit-exact vs the unfused module sequence.

242 - """

241 + launch. Bit-exact vs the unfused module sequence."""

242 ids = input_ids.view(-1)

243 (T,) = ids.shape

244 n = embed_table.shape[1]

@@ -333,8 +332,7 @@ def embed_dual_rmsnorm_cat(

332 (draft decode steps) or precomputed ``embeds`` ([T, N], the target-merged

333 multimodal embeddings at draft prefill); ``pre_norm_weight`` chains the

334 backbone embed_norm in front of the depth embed_norm (bit-exact vs the

336 - unfused sequence). The concat copies collapse into direct writes.

337 - """

335 + unfused sequence). The concat copies collapse into direct writes."""

336 T, n = hidden.shape

337 if embeds is not None:

338 assert embeds.shape == hidden.shape

vllm/models/inkling/nvidia/model.py

+1 -2

@@ -75,8 +75,7 @@ def _sconv_add_norm(

75 """``h = hidden + sconv(TP-sum(delta)); y = rmsnorm(h)``.

76

77 The Lamport path performs reduce-scatter + shard sconv + all-gather +

78 - residual add + norm. The NCCL path handles unsupported configurations.

79 - """

78 + residual add + norm. The NCCL path handles unsupported configurations."""

79 attn_metadata = get_forward_context().attn_metadata

80 m = (

81 attn_metadata.get(sconv.owner.prefix)

vllm/models/inkling/nvidia/moe.py

+3 -6

@@ -191,8 +191,7 @@ def inkling_gate_select(

191

192 class InklingGate(nn.Module):

193 """Sigmoid gate with selection bias, log-sigmoid renorm after top-k, and

194 - global scale (the served checkpoint's only configuration).

195 - """

194 + global scale (the served checkpoint's only configuration)."""

195

196 def __init__(

197 self,

@@ -244,8 +243,7 @@ class InklingGate(nn.Module):

243 self, gating_output: torch.Tensor

244 ) -> tuple[torch.Tensor, torch.Tensor]:

245 """Full selection: (weights, ids) of [T, K + S]. The first K entries

247 - are the routed top-k; the S trailing entries are the sink gammas.

248 - """

246 + are the routed top-k; the S trailing entries are the sink gammas."""

247 return inkling_gate_select(

248 gating_output,

249 self.n_total_experts,

@@ -266,8 +264,7 @@ class InklingGate(nn.Module):

264 def _inkling_moe_ep_size() -> int:

265 """EP size the FusedMoE layer will run with (mirrors

266 FusedMoEParallelConfig.make: experts shard over tp * dp * pcp when

269 - expert parallelism is enabled).

270 - """

267 + expert parallelism is enabled)."""

268 parallel_config = get_current_vllm_config().parallel_config

269 if not parallel_config.enable_expert_parallel:

270 return 1

vllm/models/inkling/nvidia/mtp.py

+3 -6

@@ -166,8 +166,7 @@ class InklingMultiTokenPredictor(nn.Module):

166 ) -> torch.Tensor:

167 """Draft-prefill embedding: fused gather + backbone embed_norm, then

168 the target's tower embeddings scattered in unnormed (the backbone

169 - convention — MM embeds are merged after embed_norm).

170 - """

169 + convention — MM embeds are merged after embed_norm)."""

170 norm = self.backbone_embed_norm

171 embeds = embed_rmsnorm(

172 input_ids,

@@ -196,8 +195,7 @@ class InklingMultiTokenPredictor(nn.Module):

195 """The depth layer's [rmsnorm(hidden) | embed_norm(embed)] input in one

196 launch: embedding row gather + the backbone embed_norm + the depth

197 embed_norm chain on one side, hidden_norm on the other, written

199 - straight into the cat buffer.

200 - """

198 + straight into the cat buffer."""

199 hidden_w = layer.hidden_norm.weight

200 embed_w = layer.embed_norm.weight

201 eps = layer.hidden_norm.variance_epsilon

@@ -329,8 +327,7 @@ class InklingMTP(nn.Module, SupportsMultiModalEmbeddings):

327 """Greedy draft tokens via rank-local argmax + tiny (value, index)

328 reduction — no full-vocab logits all-gather. The muP divisor is a

329 positive scalar, so the argmax is invariant and the scaling is

332 - skipped entirely.

333 - """

330 + skipped entirely."""

331 return self.logits_processor.get_top_tokens(self.lm_head, hidden_states)

332

333 def load_weights(self, weights: Iterable[tuple[str, torch.Tensor]]) -> set[str]:

vllm/models/inkling/nvidia/ops/mm_towers.py

+2 -4

@@ -60,8 +60,7 @@ def dmel_embed_sum_norm(

60 eps: float,

61 ) -> torch.Tensor:

62 """``rmsnorm(sum_b weight[b * VOCAB + idx[:, b]])`` in one launch (no

63 - [T, NB, D] intermediate).

64 - """

63 + [T, NB, D] intermediate)."""

64 T, nb = dmel_idx.shape

65 D = weight.shape[1]

66 assert weight.shape[0] % nb == 0

@@ -152,8 +151,7 @@ def rmsnorm_gelu(

151 ) -> torch.Tensor:

152 """Fused ``gelu(rmsnorm(x))`` (or plain rmsnorm); multiple rows per block

153 when D is small. With ``fold``, x must be [N, T, H, W, D] and the output

155 - comes back as ``fold_timespace_to_depth(result, *fold)``.

156 - """

154 + comes back as ``fold_timespace_to_depth(result, *fold)``."""

155 D = x.shape[-1]

156 flat = x.reshape(-1, D)

157 assert flat.stride(1) == 1 and flat.stride(0) == D

vllm/models/inkling/nvidia/ops/norm.py

+2 -4

@@ -238,8 +238,7 @@ def embed_rmsnorm(

238 embedding-table gather.

239 ``chain_weight`` additionally emits ``rmsnorm(out, chain_weight)`` (the

240 first decoder layer's pre-attention norm) as a second output, still one

241 - launch. Bit-exact vs the unfused module sequence.

242 - """

241 + launch. Bit-exact vs the unfused module sequence."""

242 ids = input_ids.view(-1)

243 (T,) = ids.shape

244 n = embed_table.shape[1]

@@ -333,8 +332,7 @@ def embed_dual_rmsnorm_cat(

332 (draft decode steps) or precomputed ``embeds`` ([T, N], the target-merged

333 multimodal embeddings at draft prefill); ``pre_norm_weight`` chains the

334 backbone embed_norm in front of the depth embed_norm (bit-exact vs the

336 - unfused sequence). The concat copies collapse into direct writes.

337 - """

335 + unfused sequence). The concat copies collapse into direct writes."""

336 T, n = hidden.shape

337 if embeds is not None:

338 assert embeds.shape == hidden.shape

vllm/models/kimi_k3/nvidia/mla.py

+1 -2

@@ -745,8 +745,7 @@ class MultiHeadLatentAttention(nn.Module, AttentionLayerBase):

745 slot_mapping: torch.Tensor,

746 ) -> torch.Tensor:

747 """Fused decode query-concat + latent cache insert, dispatched by cache

748 - dtype (same policy as prefill: fp8 cache -> fp8 query).

749 - """

748 + dtype (same policy as prefill: fp8 cache -> fp8 query)."""

749 if self.kv_cache_dtype == "fp8_ds_mla":

750 cache = self.kv_cache

751 if cache.dtype != torch.uint8:

vllm/models/minimax_m3/amd/model.py

+1 -2

@@ -334,8 +334,7 @@ def _aiter_moe_fused_shared_experts_enabled(

334

335 class MiniMaxM3MoE(nn.Module):

336 """Sigmoid-routed MoE block with a routing-bias correction and a shared

337 - expert.

338 - """

337 + expert."""

338

339 def __init__(

340 self,

vllm/models/minimax_m3/common/indexer.py

+2 -4

@@ -173,8 +173,7 @@ class MiniMaxM3IndexerPrefillMetadata:

173 @dataclass

174 class MiniMaxM3IndexerDecodeMetadata:

175 """Per-decode state (cudagraph-safe). ``decode_query_len`` is the uniform

176 - per-request query length (1, or 1 + num_speculative_tokens).

177 - """

176 + per-request query length (1, or 1 + num_speculative_tokens)."""

177

178 seq_lens: torch.Tensor # [num_decodes] int32

179 block_table: torch.Tensor

@@ -207,8 +206,7 @@ class MiniMaxM3IndexerMetadataBuilder(

206 AttentionMetadataBuilder[MiniMaxM3IndexerMetadata]

207 ):

208 """Abstract base: shared setup only. The Triton and MSA builders are

210 - parallel subclasses that each own their full ``build`` (no shared code).

211 - """

209 + parallel subclasses that each own their full ``build`` (no shared code)."""

210

211 # Full cudagraphs for uniform decode batches (incl. spec-decode verify).

212 _cudagraph_support: ClassVar[AttentionCGSupport] = AttentionCGSupport.UNIFORM_BATCH

vllm/models/minimax_m3/common/sparse_attention.py

+1 -2

@@ -203,8 +203,7 @@ class MiniMaxM3SparsePrefillMetadata:

203 @dataclass

204 class MiniMaxM3SparseDecodeMetadata:

205 """Per-decode state (cudagraph-safe). ``decode_query_len`` is the uniform

206 - per-request query length (1, or 1 + num_speculative_tokens).

207 - """

206 + per-request query length (1, or 1 + num_speculative_tokens)."""

207

208 seq_lens: torch.Tensor # [num_decodes] int32

209 block_table: torch.Tensor

vllm/models/minimax_m3/nvidia/indexer_msa.py

+2 -4

@@ -85,8 +85,7 @@ class MiniMaxM3IndexerMSAPrefillMetadata:

85 class MiniMaxM3IndexerMSAMetadata(MiniMaxM3IndexerMetadata):

86 """Decode reuses the inherited base ``decode`` field (the Triton decode

87 metadata); ``prefill_msa`` carries the fmha score plan for the prefill side

88 - (the base ``prefill`` field is unused on this path).

89 - """

88 + (the base ``prefill`` field is unused on this path)."""

89

90 prefill_msa: MiniMaxM3IndexerMSAPrefillMetadata | None = None

91 # Per-forward view (``[:num_tokens]``) of the builder's persistent unified

@@ -112,8 +111,7 @@ class MiniMaxM3IndexerMSAMetadata(MiniMaxM3IndexerMetadata):

111

112 class MiniMaxM3IndexerMSAMetadataBuilder(MiniMaxM3IndexerMetadataBuilder):

113 """Decode metadata is the cudagraph-safe Triton decode metadata; the prefill

115 - fmha plan is built eagerly (prefill batches are not captured).

116 - """

114 + fmha plan is built eagerly (prefill batches are not captured)."""

115

116 _cudagraph_support: ClassVar[AttentionCGSupport] = AttentionCGSupport.UNIFORM_BATCH

117

vllm/models/minimax_m3/nvidia/model.py

+2 -4

@@ -190,8 +190,7 @@ class MiniMaxM3MLP(nn.Module):

190

191 class MiniMaxM3MoE(nn.Module):

192 """Sigmoid-routed MoE block with a routing-bias correction and a shared

193 - expert.

194 - """

193 + expert."""

194

195 def __init__(

196 self,

@@ -783,8 +782,7 @@ class MiniMaxM3DecoderLayer(nn.Module):

782 @property

783 def ffn_all_reduce_deferred(self) -> bool:

784 """This layer's FFN output is left un-reduced; the caller fuses the

786 - all-reduce into the next RMSNorm.

787 - """

785 + all-reduce into the next RMSNorm."""

786 if self.is_moe_layer:

787 return self.block_sparse_moe.experts.moe_config.skip_final_all_reduce

788 return not self.mlp.down_proj.reduce_results

vllm/multimodal/cache.py

+2 -4

@@ -606,8 +606,7 @@ class ShmObjectStoreSenderCache(BaseMultiModalProcessorCache):

606 @override

607 def touch_sender_cache_item(self, mm_hash: str) -> None:

608 """Touch the item in shared memory cache to prevent eviction.

609 - Increments writer_flag on sender side.

610 - """

609 + Increments writer_flag on sender side."""

610 self._shm_cache.touch(mm_hash)

611

612 @override

@@ -822,8 +821,7 @@ class ShmObjectStoreReceiverCache(BaseMultiModalReceiverCache):

821 mm_item: MultiModalKwargsItem | None = None,

822 ) -> None:

823 """Touch the item in shared memory cache to prevent eviction.

825 - Increments reader_count on receiver side.

826 - """

824 + Increments reader_count on receiver side."""

825 assert mm_item is not None

826 if "address" in mm_item:

827 address = cast(int, mm_item["address"].data)

vllm/multimodal/image.py

+1 -2

@@ -38,8 +38,7 @@ def rgba_to_rgb(

38

39 def _has_transparency(image: Image.Image) -> bool:

40 """Detect whether an image carries transparency data (RGBA, LA, PA,

41 - or tRNS chunk in P/L/RGB PNGs).

42 - """

41 + or tRNS chunk in P/L/RGB PNGs)."""

42 if image.mode in ("RGBA", "LA", "PA"):

43 return True

44 return "transparency" in getattr(image, "info", {})

vllm/multimodal/inputs.py

+1 -2

@@ -295,8 +295,7 @@ def _is_dense(t: torch.Tensor) -> bool:

295 """Whether `t`'s elements fill a contiguous storage range, possibly

296 permuted (e.g. a transposed matrix). Such layouts can be copied with

297 pitched cudaMemcpy2D/3DAsync directly from pinned memory, whereas

298 - gapped layouts (e.g. strided slices) stage through a pageable temp.

299 - """

298 + gapped layouts (e.g. strided slices) stage through a pageable temp."""

299 if not t.is_contiguous():

300 expected_stride = 1

301 for stride, size in sorted(zip(t.stride(), t.shape)):

vllm/multimodal/utils.py

+1 -2

@@ -220,8 +220,7 @@ def strip_covered_mm_data(

220 so the workers never consume the payload fields. M-RoPE models keep

221 CPU-side metadata fields used to compute positions. SHM address items

222 are also kept so workers can balance the sender's reference count. The

223 - scheduler-side ``Request`` keeps the full features.

224 - """

223 + scheduler-side ``Request`` keeps the full features."""

224 if not mm_features or num_computed_tokens == 0:

225 return mm_features

226

vllm/parser/engine/incremental_lexer.py

+1 -2

@@ -1,8 +1,7 @@

1 # SPDX-License-Identifier: Apache-2.0

2 # SPDX-FileCopyrightText: Copyright contributors to the vLLM project

3 """Incremental text lexer that converts text chunks into terminal

4 -tokens, with prefix-match buffering for ambiguous boundaries.

5 -"""

4 +tokens, with prefix-match buffering for ambiguous boundaries."""

5

6 from __future__ import annotations

7

vllm/parser/engine/streaming_parser_engine.py

+1 -2

@@ -1,8 +1,7 @@

1 # SPDX-License-Identifier: Apache-2.0

2 # SPDX-FileCopyrightText: Copyright contributors to the vLLM project

3 """Streaming parser engine that orchestrates token ID scanning,

4 -incremental lexing, and state-machine-driven semantic event emission.

5 -"""

4 +incremental lexing, and state-machine-driven semantic event emission."""

5

6 from __future__ import annotations

7

vllm/parser/engine/token_id_scanner.py

+1 -2

@@ -1,8 +1,7 @@

1 # SPDX-License-Identifier: Apache-2.0

2 # SPDX-FileCopyrightText: Copyright contributors to the vLLM project

3 """Scan delta token IDs for special tokens and split the stream into

4 -pre-lexed terminals and plain text chunks.

5 -"""

4 +pre-lexed terminals and plain text chunks."""

5

6 from __future__ import annotations

7

vllm/parser/inkling.py

+1 -2

@@ -67,8 +67,7 @@ _WS = " \t\r\n"

67

68 def _scan_json_value(raw: str, start: int) -> int | None:

69 """Return the end index (exclusive) of the JSON object starting at

70 - ``raw[start]``, or ``None`` when the object is still unterminated.

71 - """

70 + ``raw[start]``, or ``None`` when the object is still unterminated."""

71 depth = 0

72 in_string = False

73 escape = False

vllm/platforms/cuda.py

+1 -2

@@ -62,8 +62,7 @@ def _cuda_device_count_stateless(cuda_visible_devices: str | None = None) -> int

62 CUDA_VISIBLE_DEVICES has already been set to the desired value.

63

64 # This can be removed and simply replaced with torch.cuda.get_device_count

65 - # after https://github.com/pytorch/pytorch/pull/122815 is released.

66 - """

65 + # after https://github.com/pytorch/pytorch/pull/122815 is released."""

66 # Note: cuda_visible_devices is not used, but we keep it as an argument for

67 # LRU Cache purposes.

68

vllm/platforms/interface.py

+1 -2

@@ -42,8 +42,7 @@ def set_assigned_physical_gpu_ids(ids: list[int]) -> None:

42 Idempotent: a second call with the same value is a no-op.

43 Raises RuntimeError if called again with a different value.

44

45 - This is expected to run during single-threaded worker initialization.

46 - """

45 + This is expected to run during single-threaded worker initialization."""

46 global _assigned_physical_gpu_ids

47 if _assigned_physical_gpu_ids is not None:

48 if _assigned_physical_gpu_ids != ids:

vllm/platforms/rocm.py

+2 -4

@@ -91,8 +91,7 @@ def _rocm_device_count_stateless(cuda_visible_devices: str | None = None) -> int

91 CUDA_VISIBLE_DEVICES has already been set to the desired value.

92

93 # This can be removed and simply replaced with torch.cuda.get_device_count

94 - # after https://github.com/pytorch/pytorch/pull/122815 is released.

95 - """

94 + # after https://github.com/pytorch/pytorch/pull/122815 is released."""

95 # Note: cuda_visible_devices is not used, but we keep it as an argument for

96 # LRU Cache purposes.

97

@@ -127,8 +126,7 @@ def _get_wsl_kernel_version() -> tuple[int, ...] | None:

126

127 def _sync_hip_cuda_env_vars():

128 """Ensure HIP_VISIBLE_DEVICES and CUDA_VISIBLE_DEVICES are consistent.

130 - Treats empty string as unset. Raises on genuine conflicts.

131 - """

129 + Treats empty string as unset. Raises on genuine conflicts."""

130 hip_val = os.environ.get("HIP_VISIBLE_DEVICES") or None

131 cuda_val = os.environ.get("CUDA_VISIBLE_DEVICES") or None

132

vllm/profiler/wrapper.py

+1 -2

@@ -98,8 +98,7 @@ class WorkerProfiler(ABC):

98

99 def step(self) -> None:

100 """Update the profiler state at each worker step,

101 - to handle delayed starts and max iteration limits.

102 - """

101 + to handle delayed starts and max iteration limits."""

102 if not self._active:

103 return

104

vllm/renderers/base.py

+1 -2

@@ -377,8 +377,7 @@ class BaseRenderer(ABC, Generic[_T]):

377

378 async def clear_mm_cache_async(self) -> None:

379 """Serialize clear_mm_cache through the multimodal executor to avoid

380 - races with concurrent process_inputs on the mm_processor_cache.

381 - """

380 + races with concurrent process_inputs on the mm_processor_cache."""

381 await self._clear_mm_cache_async()

382

383 def shutdown(self) -> None:

vllm/renderers/hf.py

+2 -4

@@ -103,8 +103,7 @@ _TOKENIZE_OVERRIDE_WARNING: Final[str] = (

103

104 def _ensure_prompt_embeds_placeholder_token(tokenizer: HfTokenizer) -> int:

105 """Register `PROMPT_EMBEDS_PLACEHOLDER_TOKEN` as a special token and return

106 - its token ID.

107 - """

106 + its token ID."""

107 cached = _PROMPT_EMBEDS_PLACEHOLDER_TOKEN_ID_CACHE.get(tokenizer)

108 if cached is not None:

109 return cached

@@ -193,8 +192,7 @@ def _build_mixed_prompt_embeds(

192 positions: list[tuple[int, int]],

193 ) -> tuple[torch.Tensor, list[bool]]:

194 """Build the full-length `prompt_embeds` tensor and the `is_token_ids`

196 - mask aligned to `token_ids`.

197 - """

195 + mask aligned to `token_ids`."""

196 total_len = len(token_ids)

197 hidden_size = prompt_embeds_tensors[0].shape[1]

198 dtype = prompt_embeds_tensors[0].dtype

vllm/renderers/online_derenderer.py

+1 -2

@@ -616,8 +616,7 @@ def _convert_chat_logprobs_to_completion_logprobs(

616 logprobs: ChatCompletionLogProbs,

617 ) -> CompletionLogProbs:

618 """Convert ChatCompletionLogProbs (per-token objects) to CompletionLogProbs

619 - (parallel flat lists) as required by the /v1/completions response schema.

620 - """

619 + (parallel flat lists) as required by the /v1/completions response schema."""

620 if logprobs.content is None:

621 return CompletionLogProbs()

622

vllm/sampling_params.py

+1 -2

@@ -800,8 +800,7 @@ class SamplingParams(

800 """Number of sample logprobs to return per output token, or `None` if

801 no sample logprobs were requested. Takes `logprob_token_ids` into

802 account: when `logprobs` is unset but `logprob_token_ids` is set,

803 - returns `len(logprob_token_ids)`.

804 - """

803 + returns `len(logprob_token_ids)`."""

804 if self.logprobs is not None:

805 return self.logprobs

806 return len(self.logprob_token_ids) if self.logprob_token_ids else None

vllm/tool_parsers/hermes_tool_parser.py

+1 -2

@@ -219,8 +219,7 @@ class Hermes2ProToolParser(ToolParser):

219

220 Returns a ``DeltaMessage`` containing either plain content (for

221 text preceding any tool call) or one or more ``DeltaToolCall``

222 - entries, or ``None`` if there is nothing new to send yet.

223 - """

222 + entries, or ``None`` if there is nothing new to send yet."""

223 try:

224 # Extract any content before tool calls.

225 content = self._extract_content(current_text)

vllm/tool_parsers/lfm2_tool_parser.py

+1 -2

@@ -114,8 +114,7 @@ class Lfm2ToolParser(ToolParser):

114 """Drop any orphan <|tool_call_end|> (and the preceding text) from

115 trailing content. LFM2 occasionally echoes the call body after the

116 first end token and caps it with a second end token; everything

117 - through the last such orphan is model garbage, not user content.

118 - """

117 + through the last such orphan is model garbage, not user content."""

118 last_orphan = raw_after.rfind(TOOL_CALL_END)

119 if last_orphan != -1:

120 return raw_after[last_orphan + len(TOOL_CALL_END) :]

vllm/tool_parsers/utils.py

+1 -2

@@ -426,8 +426,7 @@ def get_json_schema_from_tools(

426

427 class UnexpectedAstError(Exception):

428 """Raised when the AST structure does not match the expected

429 - pythonic tool call format.

430 - """

429 + pythonic tool call format."""

430

431 pass

432

vllm/transformers_utils/config.py

+2 -4

@@ -539,8 +539,7 @@ def register_config_parser(config_format: str):

539

540 def set_default_rope_theta(config: PretrainedConfig, default_theta: float) -> None:

541 """Some models may have no rope_theta in their config but still use RoPE.

542 - This function sets a default rope_theta if it's missing.

543 - """

542 + This function sets a default rope_theta if it's missing."""

543 if getattr(config, "rope_parameters", None) is None:

544 config.rope_parameters = {"rope_type": "default"}

545 if "rope_theta" not in config.rope_parameters:

@@ -549,8 +548,7 @@ def set_default_rope_theta(config: PretrainedConfig, default_theta: float) -> No

548

549 def patch_legacy_rope_type(rope_parameters: dict[str, Any] | None) -> None:

550 """Patch legacy RoPE type fields for backwards compatibility with

552 - older custom models which would otherwise fail to load.

553 - """

551 + older custom models which would otherwise fail to load."""

552 # No RoPE parameters to patch

553 if rope_parameters is None:

554 return

vllm/transformers_utils/configs/extract_hidden_states.py

+1 -2

@@ -1,8 +1,7 @@

1 # SPDX-License-Identifier: Apache-2.0

2 # SPDX-FileCopyrightText: Copyright contributors to the vLLM project

3 """Config definitions for ExtractHiddenStatesModel, to be used with

4 -the extract_hidden_states spec decoding method.

5 -"""

4 +the extract_hidden_states spec decoding method."""

5

6 import os

7

vllm/transformers_utils/processor.py

+3 -6

@@ -37,8 +37,7 @@ def _transformers_v4_compatibility_import():

37 subset of `ProcessorChatTemplateKwargs` as defined in Transformers v4.

38 In Transformers v5 these were merged into `ProcessorChatTemplateKwargs` and

39 `ChatTemplateLoadKwargs` was removed. For backward compatibility, we add an alias

40 - for `ChatTemplateLoadKwargs` if it doesn't exist.

41 - """

40 + for `ChatTemplateLoadKwargs` if it doesn't exist."""

41 old_import = getattr(processing_utils, "ChatTemplateLoadKwargs", None)

42 new_import = getattr(processing_utils, "ProcessorChatTemplateKwargs", None)

43 if old_import is None and new_import is not None:

@@ -53,8 +52,7 @@ def _transformers_v4_compatibility_init() -> Any:

52 processor instance before calling the original `ProcessorMixin.__init__`.

53

54 This can be removed if `Molmo2ForConditionalGeneration` is upstreamed to

56 - Transformers.

57 - """

55 + Transformers."""

56 if hasattr(ProcessorMixin.__init__, "_vllm_patched"):

57 return

58

@@ -397,8 +395,7 @@ def get_feature_extractor(

395 **kwargs: Any,

396 ):

397 """Load an audio feature extractor for the given model name

400 - via HuggingFace.

401 - """

398 + via HuggingFace."""

399 try:

400 processor_name = convert_model_repo_to_path(processor_name)

401 feature_extractor = AutoFeatureExtractor.from_pretrained(

vllm/transformers_utils/processors/cohere_asr.py

+1 -2

@@ -537,8 +537,7 @@ class CohereASRFeatureExtractor(SequenceFeatureExtractor):

537

538 class CohereASRProcessor(ProcessorMixin):

539 """HF-compatible processor combining CohereASRFeatureExtractor and a

540 - tokenizer.

541 - """

540 + tokenizer."""

541

542 feature_extractor_class = "CohereASRFeatureExtractor"

543 tokenizer_class = "AutoTokenizer"

vllm/transformers_utils/processors/glm5next.py

+2 -4

@@ -261,8 +261,7 @@ def _resize_or_pad(

261 allow_upscale: bool = False,

262 ) -> torch.Tensor:

263 """Resize onto the aligned canvas, or keep the aspect ratio and

264 - zero-pad the right/bottom sides (``resize_mode="pad"``).

265 - """

264 + zero-pad the right/bottom sides (``resize_mode="pad"``)."""

265 height, width = stacked_images.shape[-2:]

266

267 if resize_mode == "resize":

@@ -308,8 +307,7 @@ def _pixel_budget(

307 ) -> tuple[int, int]:

308 """(min_pixels, max_pixels) from the token bounds of

309 ``processor_config.json``; one vision token covers

311 - ``temporal_patch_size * (patch_size * merge_size) ** 2`` pixels.

312 - """

310 + ``temporal_patch_size * (patch_size * merge_size) ** 2`` pixels."""

311 if min_image_tokens is None or max_image_tokens is None:

312 raise ValueError(

313 "min_image_tokens and max_image_tokens must be provided by "

vllm/transformers_utils/utils.py

+1 -2

@@ -116,8 +116,7 @@ def parse_safetensors_file_metadata(path: str | PathLike) -> dict[str, Any]:

116

117 def convert_model_repo_to_path(model_repo: str) -> str:

118 """When VLLM_USE_MODELSCOPE is True convert a model

119 - repository string to a Path str.

120 - """

119 + repository string to a Path str."""

120 if not envs.VLLM_USE_MODELSCOPE or Path(model_repo).exists():

121 return model_repo

122 from modelscope.utils.file_utils import get_model_cache_root

vllm/utils/cpu_resource_utils.py

+1 -2

@@ -274,8 +274,7 @@ def get_visible_memory_node() -> list[int]:

274 def _synthesize_cpu_list() -> list[LogicalCPUInfo]:

275 """Synthesize a flat CPU list: each logical CPU is its own core on

276 NUMA node 0. Used when lscpu output is unavailable or unparsable

277 - (e.g. macOS, RISC-V).

278 - """

277 + (e.g. macOS, RISC-V)."""

278 cpu_count = os.cpu_count()

279 assert cpu_count

280 return [LogicalCPUInfo(i, i, 0) for i in range(cpu_count)]

vllm/utils/deep_gemm.py

+1 -2

@@ -704,8 +704,7 @@ def fp8_fp4_paged_mqa_logits(

704

705 def has_deep_gemm_sparse_mqa() -> bool:

706 """Whether the installed DeepGEMM provides the sparse-indexer kernels

707 - (``fp8_fp4_(paged_)sparse_mqa_logits``, added in DeepGEMM 2.8, SM100-only).

708 - """

707 + (``fp8_fp4_(paged_)sparse_mqa_logits``, added in DeepGEMM 2.8, SM100-only)."""

708 _lazy_init()

709 return (

710 _fp8_fp4_sparse_mqa_logits_impl is not None

vllm/utils/gpu_sync_debug.py

+2 -4

@@ -72,8 +72,7 @@ _sync_filter_head: tuple | None = None

72

73 def _sync_warning_hook(message, category, filename, lineno, file=None, line=None):

74 """Turn torch's sync warning into an error, but only on a thread that is

75 - being checked and is outside any allow region.

76 - """

75 + being checked and is outside any allow region."""

76 if _TORCH_SYNC_WARNING in str(message):

77 mode = _checking.get()

78 if mode is None or _allow_depth.get():

@@ -252,8 +251,7 @@ def _is_dense(t: torch.Tensor) -> bool:

251 """Whether `t`'s elements fill a contiguous storage range, possibly

252 permuted (e.g. a transposed matrix). Such layouts can be copied with

253 pitched cudaMemcpy2D/3DAsync directly from/to pinned memory, whereas

255 - gapped layouts (e.g. strided slices) stage through a pageable temp.

256 - """

254 + gapped layouts (e.g. strided slices) stage through a pageable temp."""

255 if t.is_contiguous():

256 # The common case, answered by a memoized TensorImpl flag.

257 return True

vllm/utils/platform_utils.py

+1 -2

@@ -29,8 +29,7 @@ def cuda_get_device_properties(

29 device, names: Sequence[str], init_cuda=False

30 ) -> tuple[Any, ...]:

31 """Get specified CUDA device property values without initializing CUDA in

32 - the current process.

33 - """

32 + the current process."""

33 if init_cuda or cuda_is_initialized():

34 props = torch.cuda.get_device_properties(device)

35 return tuple(getattr(props, name) for name in names)

vllm/utils/torch_utils.py

+1 -2

@@ -230,8 +230,7 @@ def _cgroup_cpu_limit() -> float | None:

230

231 def available_cpu_count() -> int:

232 """CPUs actually usable by this process: scheduling affinity capped by

233 - the cgroup CPU quota (unlike `os.cpu_count()`, which is quota-blind).

234 - """

233 + the cgroup CPU quota (unlike `os.cpu_count()`, which is quota-blind)."""

234 if sys.platform != "linux":

235 return os.cpu_count() or 1

236 count = len(os.sched_getaffinity(0))

vllm/v1/attention/backend.py

+2 -4

@@ -365,8 +365,7 @@ class AttentionBackend(ABC):

365 @classmethod

366 def supported_kv_cache_layouts(cls) -> tuple[KVCacheLayout, ...] | None:

367 """Layouts this backend's kernels can consume, most preferred first, or

368 - None when the kernels consume any layout and express no preference.

369 - """

368 + None when the kernels consume any layout and express no preference."""

369 return None

370

371 @classmethod

@@ -558,8 +557,7 @@ M = TypeVar("M")

557 class AttentionCGSupport(Enum):

558 """Constants for the cudagraph support of the attention backend

559 Here we do not consider the cascade attention, as currently

561 - it is never cudagraph supported.

562 - """

560 + it is never cudagraph supported."""

561

562 ALWAYS = 3

563 """Cudagraph always supported; supports mixed-prefill-decode"""

vllm/v1/attention/backends/cpu_attn.py

+1 -2

@@ -84,8 +84,7 @@ class CPUAttentionBackend(AttentionBackend):

84 @classmethod

85 def supports_attn_type(cls, attn_type: str) -> bool:

86 """CPU attention supports decoder,

87 - encoder-only and encoder-decoder attention.

88 - """

87 + encoder-only and encoder-decoder attention."""

88 return attn_type in (

89 AttentionType.DECODER,

90 AttentionType.ENCODER,

vllm/v1/attention/backends/flashinfer.py

+1 -2

@@ -398,8 +398,7 @@ class FlashInferBackend(AttentionBackend):

398 @classmethod

399 def customize_spec(cls, spec: "AttentionSpec") -> "AttentionSpec":

400 """NVFP4 stores K and V as separate per-head slots of packed fp4 data

401 - plus fp8 block scales.

402 - """

401 + plus fp8 block scales."""

402 if spec.state_content_bytes is not None or not spec.kv_quant_mode.is_nvfp4:

403 return spec

404 hs_k = nvfp4_kv_cache_full_dim(spec.head_size)

vllm/v1/attention/backends/mla/amx_mla.py

+1 -2

@@ -63,8 +63,7 @@ _SPLIT_OCCUPANCY_MULTIPLIER = 2

63

64 def _compute_num_kv_splits(max_seq_len: int, num_threads: int) -> int:

65 """Mirrors TritonMLAImpl's _compute_num_kv_splits, using the CPU thread

66 - count in place of SM count.

67 - """

66 + count in place of SM count."""

67 ideal_splits = 1

68 while ideal_splits < max(1, max_seq_len // _MIN_WORK_PER_SPLIT):

69 ideal_splits *= 2

vllm/v1/attention/backends/mla/flashmla_sparse.py

+1 -2

@@ -847,8 +847,7 @@ class FlashMLASparseImpl(SparseMLACommonImpl[FlashMLASparseMetadata]):

847 prefill_meta: "FlashMLASparseMetadata.FP8SeparatePrefillDecode.Prefill",

848 ) -> tuple[torch.Tensor, torch.Tensor, torch.Tensor]:

849 """All-gather this rank's upconverted KV shard so the chunk's rows attend

850 - the whole context, and map their top-k onto the rank-major result.

851 - """

850 + the whole context, and map their top-k onto the rank-major result."""

851 shard_rows = int(chunk.chunk_tot_seqlen)

852 assert self.gathered_kv_workspace is not None

853 gathered_kv = self.gathered_kv_workspace[: self.dcp_world_size * shard_rows]

vllm/v1/attention/backends/mla/indexer.py

+1 -2

@@ -1118,8 +1118,7 @@ class DeepseekV32IndexerMetadataBuilder(AttentionMetadataBuilder):

1118

1119 def _prefill_split_seq_lens(self, seq_lens_cpu: torch.Tensor) -> torch.Tensor:

1120 """Per-request KV lengths the prefill chunker budgets logits with;

1121 - subclasses whose logits rows are wider than the context override.

1122 - """

1121 + subclasses whose logits rows are wider than the context override."""

1122 return seq_lens_cpu

1123

1124 @staticmethod

vllm/v1/attention/backends/mla/prefill/base.py

+1 -2

@@ -70,8 +70,7 @@ class MLAPrefillBackend(ABC):

70 def supports_quant_output(self, quant_key: "QuantKey") -> bool:

71 """Whether `run_prefill_new_tokens` can write quantized output

72 directly (fused) for the given quant key, skipping the post-quant

73 - pass. Overridden by backends that support it.

74 - """

73 + pass. Overridden by backends that support it."""

74 return False

75

76 def supports_out(self) -> bool:

vllm/v1/attention/backends/rocm_aiter_unified_attn.py

+1 -2

@@ -110,8 +110,7 @@ class RocmAiterUnifiedAttentionBackend(RocmAttentionBackend):

110 @classmethod

111 def customize_spec(cls, spec: AttentionSpec) -> AttentionSpec:

112 """Keep K and V packed in the content dim, unlike the native HIP

113 - kernels the base class targets.

114 - """

113 + kernels the base class targets."""

114 # block_size == 1 is the per-token page-size probe (see

115 # Platform.get_page_size_bytes); real blocks must be gatherable in

116 # 16-token units by the ROCm kernel.

vllm/v1/attention/backends/rocm_attn.py

+1 -2

@@ -244,8 +244,7 @@ class RocmAttentionBackend(AttentionBackend):

244 @classmethod

245 def customize_spec(cls, spec: AttentionSpec) -> AttentionSpec:

246 """K and V as two head groups so the native HIP kernels address each side

247 - as one contiguous region (x-packed interior applied in split_kv_cache).

248 - """

247 + as one contiguous region (x-packed interior applied in split_kv_cache)."""

248 if spec.state_content_bytes is not None:

249 return spec

250 assert spec.head_size == spec.head_size_v, (

vllm/v1/attention/backends/triton_attn.py

+1 -2

@@ -275,8 +275,7 @@ class TritonAttentionBackend(AttentionBackend):

275 @classmethod

276 def customize_spec(cls, spec: "AttentionSpec") -> "AttentionSpec":

277 """Per-token-head modes pack inline fp32 scales after each head's

278 - data, so the content is (data + one scale) per K/V side.

279 - """

278 + data, so the content is (data + one scale) per K/V side."""

279 mode = spec.kv_quant_mode

280 if spec.state_content_bytes is not None or not mode.is_per_token_head:

281 return spec

vllm/v1/attention/ops/turboquant_soa/triton_turboquant_store.py

+1 -2

@@ -45,8 +45,7 @@ def _store_quantized_value(

45 BLOCK_GRP: tl.constexpr,

46 ):

47 """Uniform quantization of values to VQB bits, pack, and store with

48 - scale/zero in the per-block SoA metadata region (Opt#3 layout).

49 - """

48 + scale/zero in the per-block SoA metadata region (Opt#3 layout)."""

49 if VQB == 3:

50 val_vec = tl.load(Value_ptr + base + d_offs, mask=d_mask, other=0.0).to(

51 tl.float32

vllm/v1/core/kv_cache_coordinator.py

+1 -2

@@ -461,8 +461,7 @@ class KVCacheCoordinator(ABC):

461 ) -> tuple[tuple[list[KVCacheBlock], ...], int, int]:

462 """Returns the per-group hit blocks, the hit length, and the number of

463 ``num_uncached_common_prefix_tokens`` (a shared prefix that a

464 - sparse-retention group has not cached yet; 0 unless hybrid).

465 - """

464 + sparse-retention group has not cached yet; 0 unless hybrid)."""

465 pass

466

467 def new_step_starts(self) -> None:

vllm/v1/core/kv_cache_manager.py

+1 -2

@@ -856,8 +856,7 @@ class KVCacheManager:

856 self, request_id: str, start_token: int, end_token: int

857 ) -> list[int]:

858 """The request's block ids covering [start_token, end_token), from

859 - the groups whose new blocks are zeroed by the worker.

860 - """

859 + the groups whose new blocks are zeroed by the worker."""

860 ids: list[int] = []

861 for mgr in self.coordinator.single_type_managers:

862 if mgr.records_new_block_ids:

vllm/v1/core/sched/interface.py

+2 -4

@@ -178,8 +178,7 @@ class SchedulerInterface(ABC):

178

179 def has_unfinished_requests(self) -> bool:

180 """Returns True if there are unfinished requests in the scheduler's

181 - internal queue.

182 - """

181 + internal queue."""

182 return self.get_num_unfinished_requests() > 0

183

184 @abstractmethod

@@ -199,8 +198,7 @@ class SchedulerInterface(ABC):

198

199 def has_requests(self) -> bool:

200 """Returns True if there are unfinished requests, or finished requests

202 - not yet returned in SchedulerOutputs.

203 - """

201 + not yet returned in SchedulerOutputs."""

202 return self.has_unfinished_requests() or self.has_finished_requests()

203

204 @property

vllm/v1/core/sched/request_queue.py

+3 -6

@@ -43,8 +43,7 @@ class RequestQueue(ABC):

43 @abstractmethod

44 def prepend_requests(self, requests: "RequestQueue") -> None:

45 """Prepend all requests from another queue to the front of this

46 - queue.

47 - """

46 + queue."""

47 pass

48

49 @abstractmethod

@@ -161,16 +160,14 @@ class PriorityRequestQueue(RequestQueue):

160 """Add a request to the queue according to priority policy.

161

162 Note: In a priority queue, there is no concept of prepending to the

164 - front. Requests are ordered by (priority, arrival_time).

165 - """

163 + front. Requests are ordered by (priority, arrival_time)."""

164 self.add_request(request)

165

166 def prepend_requests(self, requests: RequestQueue) -> None:

167 """Add all requests from another queue according to priority policy.

168

169 Note: In a priority queue, there is no concept of prepending to the

172 - front. Requests are ordered by (priority, arrival_time).

173 - """

170 + front. Requests are ordered by (priority, arrival_time)."""

171 for request in requests:

172 self.add_request(request)

173

vllm/v1/engine/async_llm.py

+1 -2

@@ -871,8 +871,7 @@ class AsyncLLM(EngineClient):

871 ) -> None:

872 """Submit a pre-aborted request so the connector's request_finished

873 hook runs to free any pre-admission KV-transfer resources (e.g. NIXL

874 - prefill blocks pinned on the P node).

875 - """

874 + prefill blocks pinned on the P node)."""

875 request = EngineCoreRequest(

876 request_id=request_id,

877 prompt_token_ids=[0],

vllm/v1/engine/core.py

+3 -6

@@ -598,8 +598,7 @@ class EngineCore:

598

599 def _should_throttle_prefills(self) -> bool:

600 """Whether to defer new prefills this step (DP prefill balancing).

601 - Overridden by the DP engine core; never throttles otherwise.

602 - """

601 + Overridden by the DP engine core; never throttles otherwise."""

602 return False

603

604 def step(self) -> tuple[dict[int, EngineCoreOutputs], bool]:

@@ -1623,8 +1622,7 @@ class EngineCoreProc(EngineCore):

1622 @staticmethod

1623 def _convert_msgspec_args(method, args):

1624 """If a provided arg type doesn't match corresponding target method

1626 - arg type, try converting to msgspec object.

1627 - """

1625 + arg type, try converting to msgspec object."""

1626 if not args:

1627 return args

1628 arg_types = signature(method).parameters.values()

@@ -2006,8 +2004,7 @@ class EngineCoreProc(EngineCore):

2004

2005 class DPEngineCoreProc(EngineCoreProc):

2006 """ZMQ-wrapper for running EngineCore in background process

2009 - in a data parallel context.

2010 - """

2007 + in a data parallel context."""

2008

2009 def __init__(

2010 self,

vllm/v1/engine/core_client.py

+7 -14

@@ -248,8 +248,7 @@ class EngineCoreClient(ABC):

248

249 def dp_engines_running(self) -> bool:

250 """Returns True if data parallel engines are collectively in a

251 - running state.

252 - """

251 + running state."""

252 raise NotImplementedError

253

254 async def commit_elastic_ep(self) -> None:

@@ -446,8 +445,7 @@ class InprocClient(EngineCoreClient):

445 @dataclass

446 class BackgroundResources:

447 """Used as a finalizer for clean shutdown, avoiding

449 - circular reference back to the client object.

450 - """

448 + circular reference back to the client object."""

449

450 ctx: zmq.Context

451 # If CoreEngineProcManager, it manages local engines;

@@ -801,8 +799,7 @@ class MPClient(EngineCoreClient):

799

800 def _apply_ready_response(self, payload: bytes) -> None:

801 """Decode an EngineCoreReadyResponse and sync any post-initialization

804 - config changes (e.g. auto-fitted max_model_len) back to the frontend.

805 - """

802 + config changes (e.g. auto-fitted max_model_len) back to the frontend."""

803 vllm_config = self.vllm_config

804 response = (

805 msgspec.msgpack.decode(payload, type=EngineCoreReadyResponse)

@@ -1333,8 +1330,7 @@ class AsyncMPClient(MPClient):

1330

1331 class DPAsyncMPClient(AsyncMPClient):

1332 """Asyncio-compatible client for multi-proc, multi-engine (data parallel)

1336 - EngineCore. Assumes external load-balancing by default.

1337 - """

1333 + EngineCore. Assumes external load-balancing by default."""

1334

1335 def __init__(

1336 self,

@@ -1518,8 +1514,7 @@ class DPAsyncMPClient(AsyncMPClient):

1514

1515 class DPLBAsyncMPClient(DPAsyncMPClient):

1516 """Asyncio-compatible client for multi-proc, multi-engine (data parallel)

1521 - EngineCore. Load-balances between multiple engine processes.

1522 - """

1517 + EngineCore. Load-balances between multiple engine processes."""

1518

1519 def __init__(

1520 self,

@@ -1827,8 +1822,7 @@ class DPLBAsyncMPClient(DPAsyncMPClient):

1822 num_redundant_experts: int,

1823 ) -> None:

1824 """Prepare scale up by creating new engine cores and reconfiguring

1830 - existing ones.

1831 - """

1825 + existing ones."""

1826 self._setup_elastic_ep_reconfig_bootstrap()

1827

1828 # Phase 1: Send reconfig messages to existing engines

@@ -1920,8 +1914,7 @@ class DPLBAsyncMPClient(DPAsyncMPClient):

1914

1915 async def _commit_scale_down_elastic_ep(self, new_data_parallel_size: int) -> None:

1916 """Scale down the data parallel size by shutting down and

1923 - reconfiguring existing engine cores.

1924 - """

1917 + reconfiguring existing engine cores."""

1918 cur_data_parallel_size = len(self.core_engines)

1919

1920 self.eep_scaling_cache = ElasticScalingCache(

vllm/v1/engine/detokenizer.py

+1 -2

@@ -147,8 +147,7 @@ class BaseIncrementalDetokenizer(IncrementalDetokenizer, ABC):

147

148 def get_next_output_text(self, finished: bool, delta: bool) -> str:

149 """If delta is True, only new text since the last call to

150 - this method is returned

151 - """

150 + this method is returned"""

151 # We return the full output text if the sequence is finished.

152 buffer_length = 0 if finished else self.stop_buffer_length

153 if not delta:

vllm/v1/engine/utils.py

+1 -2

@@ -1041,8 +1041,7 @@ def get_engine_zmq_addresses(

1041

1042 Set ``defer_api_server_ports=False`` only when the consumer cannot

1043 report a bound port back (e.g. the Rust front-end). IPC paths are

1044 - unaffected.

1045 - """

1044 + unaffected."""

1045 parallel_config = vllm_config.parallel_config

1046 local_engine_count = parallel_config.data_parallel_size_local

1047 local_start_index = parallel_config.data_parallel_rank_local

vllm/v1/executor/abstract.py

+1 -2

@@ -284,8 +284,7 @@ class Executor(ABC):

284 @abstractmethod

285 def check_health(self) -> None:

286 """Checks if the executor is healthy. If not, it should raise an

287 - exception.

288 - """

287 + exception."""

288 raise NotImplementedError

289

290 def shutdown(self) -> None:

vllm/v1/executor/multiproc_executor.py

+4 -8

@@ -384,8 +384,7 @@ class MultiprocExecutor(Executor):

384 ec_output_aggregator: ECOutputAggregator | None = None,

385 ) -> Any:

386 """Returns single result if unique_reply_rank and/or an output

387 - aggregator is provided, otherwise list.

388 - """

387 + aggregator is provided, otherwise list."""

388 assert self.rpc_broadcast_mq is not None, (

389 "collective_rpc should not be called on follower node"

390 )

@@ -452,8 +451,7 @@ class MultiprocExecutor(Executor):

451 def _ensure_worker_termination(worker_procs: list[BaseProcess]):

452 """Ensure that all worker processes are terminated. Assumes workers have

453 received termination requests. Waits for processing, then sends

455 - termination and kill signals if needed.

456 - """

454 + termination and kill signals if needed."""

455

456 def wait_for_termination(procs, timeout):

457 if not time:

@@ -854,8 +852,7 @@ class WorkerProc:

852 @staticmethod

853 def worker_main(*args, **kwargs):

854 """Worker initialization and execution loops.

857 - This runs a background process

858 - """

855 + This runs a background process"""

856 # Signal handler used for graceful termination.

857 # SystemExit exception is only raised once to allow this and worker

858 # processes to terminate without error

@@ -1099,8 +1096,7 @@ class WorkerProc:

1096 def set_multiprocessing_worker_envs(local_world_size: int = 1):

1097 """Set up environment variables that should be used when there are workers

1098 in a multiprocessing environment. This should be called by the parent

1102 - process before worker processes are created

1103 - """

1099 + process before worker processes are created"""

1100 _maybe_force_spawn()

1101

1102 if current_platform.is_cpu() or "OMP_NUM_THREADS" in os.environ:

vllm/v1/executor/ray_utils.py

+1 -2

@@ -55,8 +55,7 @@ try:

55

56 class RayWorkerWrapper(WorkerWrapperBase):

57 """Ray wrapper for vllm.worker.Worker, allowing Worker to be

58 - lazily initialized after Ray sets CUDA_VISIBLE_DEVICES.

59 - """

58 + lazily initialized after Ray sets CUDA_VISIBLE_DEVICES."""

59

60 def __init__(self, *args, **kwargs) -> None:

61 super().__init__(*args, **kwargs)

vllm/v1/kv_cache_interface.py

+1 -2

@@ -1252,8 +1252,7 @@ class UniformTypeKVCacheSpecs(KVCacheSpec):

1252

1253 def get_max_layers_per_page_size(self) -> int:

1254 """Max number of layers sharing a page size. For a balanced bucket

1255 - this equals the number of repetitions of the layer pattern.

1256 - """

1255 + this equals the number of repetitions of the layer pattern."""

1256 return Counter(

1257 spec.page_size_bytes for spec in self.kv_cache_specs.values()

1258 ).most_common(1)[0][1]

vllm/v1/kv_cache_layout.py

+1 -2

@@ -49,8 +49,7 @@ class KVCacheLayout(Enum):

49 @property

50 def is_block_compact(self) -> bool:

51 """True when each page's [H, N, C] bytes form one contiguous run; i.e.

52 - the L and B dimensions are outermost.

53 - """

52 + the L and B dimensions are outermost."""

53 return set(self.value[:2]) == {_DIM_L, _DIM_B}

54

55 @property

vllm/v1/kv_offload/base.py

+2 -4

@@ -472,8 +472,7 @@ class CopyRun:

472 """A strided byte correspondence between this worker's physical page and

473 a canonical page: for i in range(num_fragments), fragment i spans

474 [local_offset + i * local_stride, +fragment_size) in the worker's page and

475 - [canonical_offset + i * canonical_stride, +fragment_size) canonically.

476 - """

475 + [canonical_offset + i * canonical_stride, +fragment_size) canonically."""

476

477 local_offset: int

478 canonical_offset: int

@@ -554,8 +553,7 @@ class TransferResult:

553 class OffloadingWorker(ABC):

554 """Runs in the worker process. Performs async KV transfers for ONE

555 offloaded medium (e.g. CPU). Direction is explicit via submit_store /

557 - submit_load, so there is no (src_medium, dst_medium) routing.

558 - """

556 + submit_load, so there is no (src_medium, dst_medium) routing."""

557

558 @abstractmethod

559 def submit_store(

vllm/v1/kv_offload/cpu/gpu_worker.py

+7 -14

@@ -126,8 +126,7 @@ def compute_sub_block_ptrs(

126 class CopyPlan(NamedTuple):

127 """Precomputed fragment-copy template for one data ref under the canonical

128 CPU layout, unrolled from the ref's mapped runs. Offsets are relative to

129 - the per-block base pointers on each side.

130 - """

129 + the per-block base pointers on each side."""

130

131 frag_offsets_src: np.ndarray

132 frag_offsets_dst: np.ndarray

@@ -165,8 +164,7 @@ def _canonical_page_ids(

164 ) -> np.ndarray:

165 """Global canonical page ids matching compute_sub_block_ptrs' enumeration.

166 These identify canonical pages consistently across ranks, so they key

168 - CanonicalPageMapping.is_writer rotation.

169 - """

167 + CanonicalPageMapping.is_writer rotation."""

168 if blocks_per_chunk == 1:

169 return block_ids[:count]

170 flat = (

@@ -179,8 +177,7 @@ def _canonical_block_sizes(

177 layer_refs_per_group: list[list[CanonicalKVCacheRef]], num_tensors: int

178 ) -> list[int]:

179 """Canonical CPU bytes per GPU block for each tensor, taken from the refs'

182 - mappings. Requires every ref to carry a mapping.

183 - """

180 + mappings. Requires every ref to carry a mapping."""

181 canonical_bytes_per_block = [0] * num_tensors

182 for layer_refs in layer_refs_per_group:

183 for ref in layer_refs:

@@ -343,8 +340,7 @@ class SingleDirectionOffloadingHandler:

340 """Upper bound on the number of copy descriptors for a transfer.

341

342 Exact for the direct layout. The canonical path may fill fewer:

346 - writer rotation later drops the blocks this rank does not write.

347 - """

343 + writer rotation later drops the blocks this rank does not write."""

344 num_copy_ops = 0

345 for g_idx, (group_size, layer_refs) in enumerate(

346 zip(group_sizes, self.layer_refs_per_group)

@@ -373,8 +369,7 @@ class SingleDirectionOffloadingHandler:

369 """Fill one group's copy descriptors for the direct (worker-private)

370 layout: one whole-page copy per (block, ref).

371

376 - Returns (op_idx past the filled descriptors, bytes added).

377 - """

372 + Returns (op_idx past the filled descriptors, bytes added)."""

373 num_bytes = 0

374 for data_ref in self.layer_refs_per_group[g_idx]:

375 t_idx = data_ref.tensor_idx

@@ -417,8 +412,7 @@ class SingleDirectionOffloadingHandler:

412 scatter each block through the ref's precomputed CopyPlan, keeping

413 only the blocks this rank writes.

414

420 - Returns (op_idx past the filled descriptors, bytes added).

421 - """

415 + Returns (op_idx past the filled descriptors, bytes added)."""

416 assert self._canonical_copy_plans is not None

417 # Zero-copy reinterpretation for pointer arithmetic: uint64 and the

418 # buffers' int64 are bit-equivalent for addresses

@@ -505,8 +499,7 @@ class SingleDirectionOffloadingHandler:

499 ) -> tuple[np.ndarray, np.ndarray]:

500 """Keep only the blocks this rank writes: replicated ranks take turns

501 writing shared canonical pages, keyed by the rank-consistent CPU-side

508 - canonical page id.

509 - """

502 + canonical page id."""

503 cpu_page_ids = _canonical_page_ids(

504 group_dst,

505 self.dst_blocks_per_chunk,

vllm/v1/kv_offload/cpu/spec.py

+2 -4

@@ -28,8 +28,7 @@ def _all_workers_barrier() -> None:

28 """Block until every worker rank has reached this point (gloo cpu group).

29

30 A superset of the node-local mmap openers suffices: once the barrier

31 - releases, every worker sharing the region file has mapped it.

32 - """

31 + releases, every worker sharing the region file has mapped it."""

32 from vllm.distributed.parallel_state import (

33 get_inner_dp_world_group,

34 get_world_group,

@@ -162,8 +161,7 @@ class CPUOffloadingSpec(OffloadingSpec):

161

162 def _uses_shared_region(self) -> bool:

163 """Whether the worker CPU buffer is the shared mmap region (vs a private

165 - per-rank tensor); replicated-layout dedup is gated on this being True.

166 - """

164 + per-rank tensor); replicated-layout dedup is gated on this being True."""

165 return current_platform.is_cuda_alike()

166

167 def create_worker(self, kv_caches: CanonicalKVCaches) -> CPUOffloadingWorker:

vllm/v1/kv_offload/tiering/async_lookup.py

+1 -2

@@ -216,8 +216,7 @@ class AsyncLookupManager(ABC):

216 def mark_miss(self, keys: Collection[OffloadKey]) -> None:

217 """Force the cached verdict for ``keys`` to False after a failed load, so

218 the scheduler stops re-issuing the doomed promotion (livelock, #49176).

219 - Keys with no cached entry are skipped.

220 - """

219 + Keys with no cached entry are skipped."""

220 for key in keys:

221 state = self._lookup_state.get(key)

222 if state is not None:

vllm/v1/kv_offload/tiering/example/manager.py

+1 -2

@@ -168,8 +168,7 @@ class ExampleSecondaryTierManager(SecondaryTierManager):

168 @override

169 def drain_jobs(self) -> None:

170 """Synchronous tier — submit_*() returns only after the operation

171 - completes, so there is nothing to wait for.

172 - """

171 + completes, so there is nothing to wait for."""

172 return

173

174 def get_num_chunks(self) -> int:

vllm/v1/kv_offload/tiering/fs/io.py

+1 -2

@@ -139,8 +139,7 @@ def _load_block(

139 ) -> None:

140 """Read one KV block from disk; remove the file only on a provable short

141 read (a too-short file is genuine corruption) and leave it untouched on any

142 - other error.

143 - """

142 + other error."""

143 fd: int | None = None

144 view_slice = view.cast("B")[offset : offset + block_size]

145 o_direct = O_DIRECT if use_o_direct else 0

vllm/v1/kv_offload/tiering/fs/manager.py

+1 -2

@@ -269,8 +269,7 @@ class FileSystemTierManager(SecondaryTierManager):

269 @override

270 def get_finished_jobs(self) -> Iterable[JobResult]:

271 """Collect finished jobs; a failed promotion marks only its failed keys

272 - as a miss here (scheduler thread).

273 - """

272 + as a miss here (scheduler thread)."""

273 results = []

274 for job_id, success, transfer_time in self._pool.get_finished():

275 if self.events is not None:

vllm/v1/kv_offload/tiering/manager.py

+1 -2

@@ -143,8 +143,7 @@ class CPUPrimaryTierOffloadingManager(CPUOffloadingManager):

143

144 class _SecondaryTierFacingParent(ParentManager):

145 """Wrapper that implements ParentManager by delegating to the

146 - TieringOffloadingManager with exclude_tier_idx set to the origin tier.

147 - """

146 + TieringOffloadingManager with exclude_tier_idx set to the origin tier."""

147

148 __slots__ = ("_m", "_origin_idx")

149

vllm/v1/kv_offload/tiering/obj/manager.py

+2 -4

@@ -308,8 +308,7 @@ class ObjectStoreSecondaryTierManager(SecondaryTierManager):

308

309 def _poll_active_transfers(self) -> None:

310 """Poll all in-flight transfers once; move newly-completed (success or

311 - failure) into ``_pending_results`` and release their NIXL handles.

312 - """

311 + failure) into ``_pending_results`` and release their NIXL handles."""

312 for job_id, entry in list(self._transfers.items()):

313 try:

314 state = self._agent.check_xfer_state(entry.xfer_handle)

@@ -364,8 +363,7 @@ class ObjectStoreSecondaryTierManager(SecondaryTierManager):

363

364 def get_finished_jobs(self) -> Iterable[JobResult]:

365 """Poll transfers; a failed promotion marks its cached verdicts False

367 - here (scheduler thread).

368 - """

366 + here (scheduler thread)."""

367 self._poll_active_transfers()

368 results = self._pending_results

369 self._pending_results = []

vllm/v1/kv_offload/tiering/p2p/manager.py

+2 -4

@@ -134,8 +134,7 @@ class P2PDestInfo:

134

135 def _parse_source(kv_params: dict | None) -> P2PSourceInfo | None:

136 """Parse the consumer sub-dict (PD ``remote_prefiller`` or symmetric

137 - ``remote_kv_source``) into a ``P2PSourceInfo``, or None if absent/incomplete.

138 - """

137 + ``remote_kv_source``) into a ``P2PSourceInfo``, or None if absent/incomplete."""

138 role = _remote_prefiller_params(kv_params)

139 do_probe = False

140 if role is None:

@@ -156,8 +155,7 @@ def _parse_source(kv_params: dict | None) -> P2PSourceInfo | None:

155

156 def _parse_dest(kv_params: dict | None) -> P2PDestInfo | None:

157 """Parse the producer ``remote_decoder`` sub-dict into a ``P2PDestInfo``,

159 - or None if the block is absent (not a remote-decode request).

160 - """

158 + or None if the block is absent (not a remote-decode request)."""

159 role = _remote_decoder_params(kv_params)

160 if role is None:

161 return None

vllm/v1/kv_offload/tiering/spec.py

+1 -2

@@ -414,8 +414,7 @@ class TieringOffloadingSpec(CPUOffloadingSpec):

414 """Require a mapping on every ref and record layer portability.

415

416 Fails loudly rather than persist direct-layout bytes under a

417 - canonical format identity.

418 - """

417 + canonical format identity."""

418 all_refs = [

419 ref for group_refs in kv_caches.group_data_refs for ref in group_refs

420 ]

vllm/v1/metrics/loggers.py

+1 -2

@@ -91,8 +91,7 @@ def load_stat_logger_plugin_factories() -> list[StatLoggerFactory]:

91

92 class AggregateStatLoggerBase(StatLoggerBase):

93 """Abstract base class for loggers that

94 - aggregate across multiple DP engines.

95 - """

94 + aggregate across multiple DP engines."""

95

96 @abstractmethod

97 def __init__(self, vllm_config: VllmConfig, engine_indexes: list[int]): ...

vllm/v1/metrics/perf.py

+1 -2

@@ -1820,8 +1820,7 @@ def get_required(obj: object, attr: str):

1820

1821 def getattr_from_list(obj: object, attrs: list[str], default: object = None):

1822 """Try to get the first attr that exists in the object

1823 - from a list of attrs. Otherwise return None.

1824 - """

1823 + from a list of attrs. Otherwise return None."""

1824 for attr in attrs:

1825 if hasattr(obj, attr):

1826 return getattr(obj, attr)

vllm/v1/metrics/ray_wrappers.py

+3 -6

@@ -82,8 +82,7 @@ class RayPrometheusMetric:

82

83 class RayGaugeWrapper(RayPrometheusMetric):

84 """Wraps around ray.util.metrics.Gauge to provide same API as

85 - prometheus_client.Gauge

86 - """

85 + prometheus_client.Gauge"""

86

87 def __init__(

88 self,

@@ -117,8 +116,7 @@ class RayGaugeWrapper(RayPrometheusMetric):

116

117 class RayCounterWrapper(RayPrometheusMetric):

118 """Wraps around ray.util.metrics.Counter to provide same API as

120 - prometheus_client.Counter

121 - """

119 + prometheus_client.Counter"""

120

121 def __init__(

122 self,

@@ -143,8 +141,7 @@ class RayCounterWrapper(RayPrometheusMetric):

141

142 class RayHistogramWrapper(RayPrometheusMetric):

143 """Wraps around ray.util.metrics.Histogram to provide same API as

146 - prometheus_client.Histogram

147 - """

144 + prometheus_client.Histogram"""

145

146 def __init__(

147 self,

vllm/v1/sample/logits_processor/builtin.py

+2 -4

@@ -129,8 +129,7 @@ class LogitBiasLogitsProcessor(LogitsProcessor):

129

130 def is_argmax_invariant(self) -> bool:

131 """Logit bias can rebalance token probabilities and change the

132 - outcome of argmax in greedy sampling.

133 - """

132 + outcome of argmax in greedy sampling."""

133 return False

134

135 def update_state(self, batch_update: BatchUpdate | None):

@@ -189,8 +188,7 @@ class MinTokensLogitsProcessor(LogitsProcessor):

188

189 def is_argmax_invariant(self) -> bool:

190 """By censoring stop tokens, min-tokens can change the outcome

192 - of the argmax operation in greedy sampling.

193 - """

191 + of the argmax operation in greedy sampling."""

192 return False

193

194 @staticmethod

vllm/v1/sample/logits_processor/state.py

+1 -2

@@ -69,8 +69,7 @@ class BatchUpdateBuilder:

69 @property

70 def removed(self) -> list[RemovedRequest]:

71 """Removed request indices sorted in

72 - descending order

73 - """

72 + descending order"""

73 self._ensure_removed_sorted()

74 return self._removed

75

vllm/v1/serial_utils.py

+2 -4

@@ -90,8 +90,7 @@ def _typestr(val: Any) -> tuple[str, str] | None:

90

91 def _encode_type_info_recursive(obj: Any) -> Any:

92 """Recursively encode type information for nested structures of

93 - lists/dicts.

94 - """

93 + lists/dicts."""

94 if obj is None:

95 return None

96 if type(obj) is list:

@@ -105,8 +104,7 @@ def _decode_type_info_recursive(

104 type_info: Any, data: Any, convert_fn: Callable[[Sequence[str], Any], Any]

105 ) -> Any:

106 """Recursively decode type information for nested structures of

108 - lists/dicts.

109 - """

107 + lists/dicts."""

108 if type_info is None:

109 return data

110 if isinstance(type_info, dict):

vllm/v1/simple_kv_offload/manager.py

+3 -6

@@ -259,8 +259,7 @@ class SimpleCPUOffloadScheduler:

259 gpu_config: "KVCacheConfig", cpu_capacity_bytes: int

260 ) -> "KVCacheConfig":

261 """Derive a CPU KVCacheConfig from the GPU config.

262 - Same kv_cache_groups, num_blocks scaled by CPU/GPU memory ratio.

263 - """

262 + Same kv_cache_groups, num_blocks scaled by CPU/GPU memory ratio."""

263 # Import here to avoid potential circular imports

264 from vllm.v1.kv_cache_interface import KVCacheTensor

265

@@ -315,8 +314,7 @@ class SimpleCPUOffloadScheduler:

314

315 def bind_gpu_block_pool(self, gpu_block_pool: BlockPool) -> None:

316 """Bind GPU block pool so that we can touch blocks during stores.

318 - Called by Scheduler after kv_cache_manager is ready.

319 - """

317 + Called by Scheduler after kv_cache_manager is ready."""

318 self._gpu_block_pool = gpu_block_pool

319

320 def get_num_new_matched_tokens(

@@ -1131,8 +1129,7 @@ class SimpleCPUOffloadScheduler:

1129 block_ids: list[int],

1130 ) -> tuple[bool, dict[str, Any] | None]:

1131 """Always returns (False, None). GPU blocks are protected by ref_cnt,

1134 - so the scheduler can free blocks immediately.

1135 - """

1132 + so the scheduler can free blocks immediately."""

1133 req_id = request.request_id

1134

1135 # Release any temp CPU hit pin from get_num_new_matched_tokens()

vllm/v1/spec_decode/extract_hidden_states.py

+1 -2

@@ -396,8 +396,7 @@ class ExtractHiddenStatesProposer:

396

397 def validate_same_kv_cache_group(self, kv_cache_config: KVCacheConfig) -> None:

398 """Validate all drafting layers belong to the same KV cache group

399 - and record the group index for common_attn_metadata selection.

400 - """

399 + and record the group index for common_attn_metadata selection."""

400 assert len(self.attn_layer_names) == 1

401 layer = self.attn_layer_names[0]

402 for gid, group in enumerate(kv_cache_config.kv_cache_groups):

vllm/v1/spec_decode/gemma4.py

+2 -4

@@ -221,8 +221,7 @@ class Gemma4Proposer(SpecDecodeBaseProposer):

221 def validate_same_kv_cache_group(self, kv_cache_config: KVCacheConfig) -> None:

222 """Draft layers span multiple KV cache groups (sliding + full

223 attention with different head dimensions), so skip the base

224 - class single-group assertion.

225 - """

224 + class single-group assertion."""

225

226 def initialize_attn_backend(

227 self,

@@ -230,8 +229,7 @@ class Gemma4Proposer(SpecDecodeBaseProposer):

229 kernel_block_sizes: list[int] | None = None,

230 ) -> None:

231 """Create separate AttentionGroup objects per KV cache spec

233 - so that each head-dim variant gets its own metadata builder.

234 - """

232 + so that each head-dim variant gets its own metadata builder."""

233 all_attn_layers = get_layers_from_vllm_config(

234 self.vllm_config,

235 AttentionLayerBase, # type: ignore[type-abstract]

vllm/v1/spec_decode/llm_base_proposer.py

+1 -2

@@ -785,8 +785,7 @@ class SpecDecodeBaseProposer:

785 block_size: int,

786 ) -> torch.Tensor:

787 """Update positions, slot mappings, and sequence metadata for the

788 - next draft step. Returns the updated positions tensor.

789 - """

788 + next draft step. Returns the updated positions tensor."""

789 positions_1d = positions[0] if self.uses_mrope else positions

790 if self.uses_mrope:

791 out_pos = self.mrope_positions[0, :batch_size]

vllm/v1/spec_decode/utils.py

+1 -2

@@ -847,6 +847,5 @@ def update_num_computed_tokens_for_batch_change(

847

848 def unconditional_to_conditional_rates(rates: list[float]) -> list[float]:

849 """Convert per-position unconditional rates to per-position conditional

850 - rates for the early-terminating rejection loop (c_i = p_i / p_{i-1}).

851 - """

850 + rates for the early-terminating rejection loop (c_i = p_i / p_{i-1})."""

851 return [p / q if q > 0.0 else 0.0 for p, q in zip(rates, [1.0, *rates[:-1]])]

vllm/v1/utils.py

+4 -8

@@ -144,8 +144,7 @@ class CpuGpuBuffer:

144

145 def copy_to_cpu(self, n: int | None = None) -> torch.Tensor:

146 """NOTE: Because this method is non-blocking, explicit synchronization

147 - is needed to ensure the data is copied to CPU.

148 - """

147 + is needed to ensure the data is copied to CPU."""

148 if n is None:

149 return self.cpu.copy_(self.gpu, non_blocking=True)

150 return self.cpu[:n].copy_(self.gpu[:n], non_blocking=True)

@@ -159,8 +158,7 @@ def get_engine_client_zmq_addr(

158 """Return an IPC path (``local_only=True``) or ``tcp://host:port``.

159

160 ``port=0`` lets the kernel assign the port at ``bind()`` time; the

162 - caller must recover it via ``getsockopt(zmq.LAST_ENDPOINT)``.

163 - """

161 + caller must recover it via ``getsockopt(zmq.LAST_ENDPOINT)``."""

162 if local_only:

163 return get_open_zmq_ipc_path()

164 return get_tcp_uri(host, port)

@@ -265,8 +263,7 @@ class APIServerProcessManager:

263 ) -> tuple[list[str], list[str]]:

264 """Return (inputs, outputs) reported by each child, indexed by

265 ``client_index``. Raises ``RuntimeError`` on timeout or premature

268 - child exit.

269 - """

266 + child exit."""

267 n = len(self._address_pipes)

268 inputs: list[str | None] = [None] * n

269 outputs: list[str | None] = [None] * n

@@ -427,8 +424,7 @@ class RustFrontendProcessManager:

424

425 class _SubprocessWrapper:

426 """Wraps subprocess.Popen to provide the BaseProcess-like interface

430 - needed by wait_for_completion_or_failure.

431 - """

427 + needed by wait_for_completion_or_failure."""

428

429 def __init__(self, proc, name: str):

430 self._proc = proc

vllm/v1/worker/encoder_cudagraph.py

+1 -2

@@ -76,8 +76,7 @@ class EncoderCudaGraphManager:

76 model: SupportsEncoderCudaGraph,

77 ):

78 """Initialize CUDA graph manager with provided token budgets

79 - and max batch size.

80 - """

79 + and max batch size."""

80 self.vllm_config = vllm_config

81 self.device = device

82 self.dtype = dtype

vllm/v1/worker/gpu/cudagraph_utils.py

+4 -8

@@ -62,8 +62,7 @@ class AttentionState(NamedTuple):

62 @dataclass(frozen=True)

63 class BatchExecutionDescriptor:

64 """Describes the shape of the batch and CG mode to run; this is used to make shape

65 - matches between the capture and runtime.

66 - """

65 + matches between the capture and runtime."""

66

67 cg_mode: CUDAGraphMode

68 num_tokens: int

@@ -91,8 +90,7 @@ def make_cudagraph_stats(

90 class CreateForwardFn(Protocol):

91 """Factory that prepares inputs (OUTSIDE the graph) and returns a

92 forward_fn. Called with warmup=True for the warmup pass and warmup=False

94 - for the captured pass.

95 - """

93 + for the captured pass."""

94

95 def __call__(

96 self,

@@ -955,8 +953,7 @@ def profile_cudagraph_memory(runner: "GPUModelRunner") -> int:

953 def _extrapolate_full_graph_memory(mem_samples: list[int], total_graphs: int) -> int:

954 """Extrapolate the total FULL capture cost from samples of the largest

955 graphs. The first capture allocates the pool baseline; later graphs mostly

958 - reuse it, so the second sample is taken as the per-graph cost.

959 - """

956 + reuse it, so the second sample is taken as the per-graph cost."""

957 if not mem_samples:

958 return 0

959 first_capture = mem_samples[0]

@@ -993,8 +990,7 @@ def _init_minimal_kv_cache_for_profiling(runner: "GPUModelRunner") -> None:

990

991 def _teardown_profiling_state(runner: "GPUModelRunner") -> None:

992 """Release the profiling KV cache and captured graphs while keeping model

996 - weights, so the real ``initialize_kv_cache`` starts from a clean slate.

997 - """

993 + weights, so the real ``initialize_kv_cache`` starts from a clean slate."""

994 torch.accelerator.synchronize()

995 if hasattr(runner.model_state, "_mamba_ctx"):

996 runner.model_state._mamba_ctx = None

vllm/v1/worker/gpu/model_runner.py

+1 -2

@@ -2228,8 +2228,7 @@ class GPUModelRunner(LoRAModelRunnerMixin):

2228

2229 def shutdown(self) -> None:

2230 """Release GPU tensors (model weights, KV caches, workspace) so that

2231 - memory is reclaimable when running in the same process.

2232 - """

2231 + memory is reclaimable when running in the same process."""

2232 torch.accelerator.synchronize()

2233 self.cudagraph_manager = None

2234 self.fast_prefill = None

vllm/v1/worker/gpu/model_states/interface.py

+1 -2

@@ -144,8 +144,7 @@ class ModelState(ABC):

144 ) -> None:

145 """Hook run on real batches before the forward pass (after block tables

146 are gathered). Used by mamba "align" prefix caching to pre-copy state

147 - across block boundaries. No-op by default.

148 - """

147 + across block boundaries. No-op by default."""

148 return None

149

150 def postprocess_state(

vllm/v1/worker/gpu/pp_utils.py

+2 -4

@@ -38,8 +38,7 @@ def compute_need_sampled_mask(input_batch: InputBatch) -> np.ndarray | None:

38 """Return a bool array of shape `[input_batch.num_reqs]` marking requests

39 that produce a sampled token this step, and therefore must have that token

40 (and the draft block proposed from it) propagated to the earlier PP stages.

41 - Returns None if no request in the batch produces a sample.

42 - """

41 + Returns None if no request in the batch produces a sample."""

42 old_computed = input_batch.num_computed_tokens_np

43 prefill_len = input_batch.prefill_len_np

44 # Exclude non-final prefill chunks (they don't produce a sample).

@@ -179,8 +178,7 @@ class PPHandler:

178

179 def receive(self, input_batch: InputBatch) -> bool:

180 """Returns True iff sampled tokens need to be gathered from *all*

182 - requests in the batch.

183 - """

181 + requests in the batch."""

182 assert not self.is_last_rank

183 need_sampled_mask = compute_need_sampled_mask(input_batch)

184 if need_sampled_mask is None:

vllm/v1/worker/gpu/sample/output.py

+1 -2

@@ -76,8 +76,7 @@ MAX_COMPACT_SUPPORT = 2048

76

77 class SamplingMaskTensors(NamedTuple):

78 """Device-side masks pending async D2H: compact ids, plus the bitmask as

79 - the exact fallback for rows wider than ``max_num_kept``.

80 - """

79 + the exact fallback for rows wider than ``max_num_kept``."""

80

81 # [num_requests, max_num_kept]

82 token_ids: torch.Tensor

vllm/v1/worker/gpu/sample/sampler.py

+1 -2

@@ -107,8 +107,7 @@ class Sampler:

107 self, idx_mapping_np: np.ndarray, include_token_ids: bool = True

108 ) -> tuple[int, int] | None:

109 """(num_logprobs, max_per_req_token_ids) for the given requests, or

110 - None when none of them want logprobs.

111 - """

110 + None when none of them want logprobs."""

111 max_num_logprobs = self.sampling_states.max_num_logprobs(idx_mapping_np)

112 max_token_ids = (

113 self.logprob_token_ids_state.max_num_token_ids(idx_mapping_np)

vllm/v1/worker/gpu/spec_decode/adaptive_verification.py

+2 -4

@@ -190,8 +190,7 @@ class AdaptiveVerificationManager:

190 """Dummy-run kwargs whose step timings seed the cost tables.

191

192 Run these inside StepTimingCollector.collect(), then hand the block's

193 - timings to set_initial_cost_curves.

194 - """

193 + timings to set_initial_cost_curves."""

194 max_num_tokens = self.req_states.max_num_batched_tokens

195 size = self._cudagraph_limit = capture_sizes[-1] if capture_sizes else 0

196 # Also profile beyond the capture limit: real steps run there

@@ -261,8 +260,7 @@ class AdaptiveVerificationManager:

260 input_batch: "InputBatch",

261 ) -> None:

262 """Publish this step's raw confidences for the ranking kernel and start

264 - copying them to the CPU, where a later step's budget reads them.

265 - """

263 + copying them to the CPU, where a later step's budget reads them."""

264 num_reqs = input_batch.num_reqs

265 ready_idx = self._stale_idx ^ 1

266 with gpu_sync_allowed():

vllm/v1/worker/gpu/spec_decode/dflash/cudagraph.py

+1 -2

@@ -72,8 +72,7 @@ def _prepare_dflash_inputs_to_capture(

72

73 class DFlashCudaGraphManager(CudaGraphManager):

74 """DFlash CudaGraphManager for the parallel-drafting query forward,

75 - building its own attention metadata from scratch.

76 - """

75 + building its own attention metadata from scratch."""

76

77 def capture(

78 self,

vllm/v1/worker/gpu/spec_decode/eagle/utils.py

+1 -2

@@ -28,8 +28,7 @@ def _should_share(eagle: nn.Module, flag: str, draft, target) -> bool:

28

29 def get_target_lm_head(target_model: nn.Module, target_language_model: nn.Module):

30 """The target's lm_head — from get_language_model() for

31 - *ForConditionalGeneration targets, else the top-level module.

32 - """

31 + *ForConditionalGeneration targets, else the top-level module."""

32 return getattr(target_language_model, "lm_head", None) or getattr(

33 target_model, "lm_head", None

34 )

vllm/v1/worker/gpu/spec_decode/speculator.py

+1 -2

@@ -268,8 +268,7 @@ class DraftModelSpeculator(BaseSpeculator):

268 @property

269 def attn_vllm_config(self) -> VllmConfig:

270 """Config for the draft's attention metadata builders. Overridden by

271 - speculators whose attention mode differs from the target's.

272 - """

271 + speculators whose attention mode differs from the target's."""

272 return self.vllm_config

273

274 def set_attn(

vllm/v1/worker/gpu_model_runner.py

+5 -10

@@ -481,8 +481,7 @@ class AsyncGPUPoolingModelRunnerOutput(AsyncModelRunnerOutput):

481

482 class ExecuteModelState(NamedTuple):

483 """Ephemeral cached state transferred between execute_model() and

484 - sample_tokens(), after execute_model() returns None.

485 - """

484 + sample_tokens(), after execute_model() returns None."""

485

486 scheduler_output: "SchedulerOutput"

487 logits: torch.Tensor

@@ -1072,8 +1071,7 @@ class GPUModelRunner(

1071 first scheduled token (column j is position start - 1 - j); -1 where

1072 the position is before the prompt or already past it. Generated

1073 positions are left to the model: under async scheduling the CPU token

1075 - table holds placeholders for them.

1076 - """

1074 + table holds placeholders for them."""

1075 buf = self.lookback_token_ids

1076 assert buf is not None

1077 buf.np.fill(-1)

@@ -3206,8 +3204,7 @@ class GPUModelRunner(

3204

3205 def _get_encoder_output_from_cache(self, mm_hash: str) -> torch.Tensor | None:

3206 """Return a cached encoder output for multimodal

3209 - embedding gather.

3210 - """

3207 + embedding gather."""

3208 return self.encoder_cache.get(mm_hash, None)

3209

3210 def _gather_mm_embeddings(

@@ -4179,8 +4176,7 @@ class GPUModelRunner(

4176

4177 This is true when `discard_request_mask` is set for every scheduled

4178 request (e.g., for chunked prefill requests that are not the last

4182 - prefill chunk).

4183 - """

4179 + prefill chunk)."""

4180 num_reqs = self.input_batch.num_reqs

4181 return bool(self.discard_request_mask.np[:num_reqs].all())

4182

@@ -6573,8 +6569,7 @@ class GPUModelRunner(

6569

6570 def shutdown(self) -> None:

6571 """Release GPU tensors (model weights, KV caches, workspace) so that

6576 - memory is reclaimable when running in the same process.

6577 - """

6572 + memory is reclaimable when running in the same process."""

6573 from vllm.model_executor.layers.rotary_embedding import _ROPE_DICT

6574 from vllm.v1.worker.workspace import reset_workspace_manager

6575

vllm/v1/worker/gpu_worker.py

+1 -2

@@ -336,8 +336,7 @@ class Worker(WorkerBase):

336 def _scoped_allocator_max_split(self, max_split_size_mb: int):

337 """Temporarily set max_split_size_mb to reduce allocator fragmentation at the

338 cost of more cudaMalloc calls (negligible in practice). Restores the original

339 - value on exit.

340 - """

339 + value on exit."""

340 if not current_platform.is_cuda():

341 yield

342 return

vllm/v1/worker/startup_plan.py

+3 -6

@@ -77,8 +77,7 @@ def compute_plan_fingerprint(

77 def _plan_path(fingerprint: str) -> str:

78 """Plans are regenerable derived state, so they live under the standard

79 vLLM cache root (like the torch.compile cache) and relocate with

80 - ``VLLM_CACHE_ROOT`` instead of needing a location knob of their own.

81 - """

80 + ``VLLM_CACHE_ROOT`` instead of needing a location knob of their own."""

81 # VLLM_CACHE_ROOT is already user-expanded by envs.py.

82 return os.path.join(

83 envs.VLLM_CACHE_ROOT, "startup_plan", f"startup_plan_{fingerprint}.json"

@@ -135,8 +134,7 @@ def _applicable_kv_cache_memory_bytes(

134 def maybe_apply_startup_plan(worker: "Worker") -> None:

135 """If enabled and ``--kv-cache-memory`` was not set explicitly, apply a

136 persisted plan by setting ``worker.cache_config.kv_cache_memory_bytes``.

138 - No-op unless ``VLLM_ENABLE_STARTUP_PLAN=1``.

139 - """

137 + No-op unless ``VLLM_ENABLE_STARTUP_PLAN=1``."""

138 if (

139 not envs.VLLM_ENABLE_STARTUP_PLAN

140 or worker.cache_config.kv_cache_memory_bytes is not None

@@ -169,8 +167,7 @@ def maybe_apply_startup_plan(worker: "Worker") -> None:

167 def maybe_save_startup_plan(worker: "Worker", kv_cache_memory_bytes: int) -> None:

168 """Atomically persist this boot's profiling result for future boots.

169 No-op unless ``VLLM_ENABLE_STARTUP_PLAN=1``; failures are logged,

172 - never raised.

173 - """

170 + never raised."""

171 if not envs.VLLM_ENABLE_STARTUP_PLAN:

172 return

173 fingerprint = compute_plan_fingerprint(

vllm/v1/worker/worker_base.py

+1 -2

@@ -128,8 +128,7 @@ class WorkerBase:

128

129 def synchronize_device(self) -> None:

130 """Block until in-flight device work completes; backends outside

131 - ``torch.accelerator`` must override with their own wait.

132 - """

131 + ``torch.accelerator`` must override with their own wait."""

132 if torch.accelerator.is_available():

133 torch.accelerator.synchronize()

134

[CI] Ignore ruff D209, rejoin the docstrings it split, and silence incompatible-rule warnings (#57212)