100 Commits
Author SHA1 Message Date
Woosuk Kwon 517e74a964 [DSV4] Refactor RoPE initialization (#44262)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-06-02 01:26:58 +00:00
Woosuk Kwon 8c3cc98cff [DSV4] Remove unncessary classes & functions (#44246)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-06-01 14:43:00 -07:00
Woosuk Kwon 27fa5aa3b9 [MRV2] Support breakable CUDA graph (#44050)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-05-30 09:40:52 -07:00
Woosuk Kwon 7bd45da585 [DSv4] Move mHC tilelang kernels & Don't use CustomOP in dsv4/nvidia (#43905)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-05-29 10:25:02 +08:00
Woosuk Kwon 69b8956dcd [Model Refactoring] Remove unncessary torch op registration for DSv4 (#43891)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-05-28 14:04:55 -07:00
Woosuk Kwon 9957e4d240 [Model Refactoring] Remove torch compile dependency in DSv4 (#43746)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-05-28 14:26:25 +00:00
Woosuk Kwon a04afd76aa [DSV4] Remove AMD/XPU path in deepseek_v4/nvidia (#43829)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-05-28 08:00:52 +00:00
Woosuk Kwon adaa5e455a [DSv4] Refactor compressor & Fix ROCm compatibility (#43710)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-05-26 19:56:46 -07:00
Woosuk Kwon 193ce8812e [DSv4] Drop _get_compressed_kv_buffer in DeepseekCompressor (#43690)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-05-26 10:11:25 -07:00
Woosuk KwonandMark McLoughlin 3aea37d28e [Doc] Add line limit to AGENTS.md (#43635)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
Signed-off-by: Mark McLoughlin <markmc@redhat.com>
Co-authored-by: Mark McLoughlin <markmc@redhat.com>
2026-05-26 09:31:23 -07:00
Woosuk Kwon c8414a8271 [ROCm] Remove MegaMoE integration in deepseek v4 (#43629)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-05-26 08:56:04 -07:00
Woosuk Kwon aa2b56ffb0 [DeepSeek V4] Move MegaMoE input prep kernel to nvidia/ops (#43632)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-05-25 21:08:29 -07:00
Woosuk Kwon 07beaed842 [Model Refactoring] Rename deepseek_v4.py to model.py [4/N] (#43077)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-05-19 01:12:46 -07:00
Woosuk Kwon b14be81c1f [Model Refactoring] Move deepseek_v4_ops to models/deepseek_v4 [3/N] (#43073)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-05-19 00:52:54 -07:00
Woosuk Kwon 87b08c5f64 [Model Refactoring] Move DeepSeek V4 layers to models/deepseek_v4/ [2/N] (#43039)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-05-18 21:00:58 -07:00
Woosuk Kwon 287471b994 [Model Refactoring] Migrate DeepSeek V4 to vllm/models/ [1/N] (#43004)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-05-18 19:50:02 -07:00
Woosuk Kwon e6ff3e9c83 [MRV2] Add shutdown() method (#41297)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-05-02 21:06:30 -07:00
Woosuk Kwon 9c61864bf8 [DeepSeek] Use torch.mm for bf16xbf16->fp32 gemm (#41300)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-04-30 16:28:57 -07:00
Woosuk Kwon 7d3195ea9f [Bugfix] Fix IMA in DSA + MTP (#40772)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-04-24 01:40:20 -07:00
Woosuk Kwon f186cfe75e [MRV2] Fix hanging issue with DeepSeek V3.2 by setting skip_attn=False (#39098)
Signed-off-by: WoosukKwon <woosuk.kwon@berkeley.edu>
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-04-06 12:55:13 -07:00
Woosuk KwonandNick Hill 144030c84e Relocate Encoder CUDA graph manager (#38116)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
Signed-off-by: Nick Hill <nickhill123@gmail.com>
Co-authored-by: Nick Hill <nickhill123@gmail.com>
2026-03-25 20:52:12 -07:00
Woosuk Kwon 4b53740d7f [MRV2] Fix for DS v3.2 (#38030)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-03-24 14:03:24 -07:00
Woosuk KwonandNick Hill ffb5b32b5f [MRV2] Consider spec decoding in warmup (#37812)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
Signed-off-by: Nick Hill <nickhill123@gmail.com>
Co-authored-by: Nick Hill <nickhill123@gmail.com>
2026-03-23 17:45:43 +00:00
Woosuk Kwon 43877a620b [MRV2] Enable PP CUDA graph test (#37830)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-03-22 16:30:25 -07:00
Woosuk Kwon a5e9d511de [MRV2] Use FP64 for Gumbel noise (#37798)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-03-22 12:28:10 -07:00
Woosuk Kwon ce9b1d76cf [MRV2] Skip hidden states allocation for PW CUDA graphs (#37818)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-03-22 11:47:21 -07:00
Woosuk Kwon e80cfe575d [MRV2] Avoid recompilation of _gather_block_tables_kernel (#37645)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-03-20 10:31:45 -07:00
Woosuk Kwon 40b8363b45 [MRV2] Use fp32 for draft logits (#37526)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-03-19 08:41:21 -07:00
Woosuk Kwon 5bc1da147f [LoRA][BugFix] Fix skipped LoRA adapters for Mistral3 (#36928)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-03-18 22:34:19 +00:00
Woosuk Kwon 96efb91480 [Model Runner V2] Fix processed logits in sample() (#37144)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-03-16 00:35:49 -07:00
Woosuk Kwon 2f8b4ce0c0 [Model Runner V2] Do not initialize sampler for non-last PP ranks (#36824)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-03-12 03:55:28 +00:00
Woosuk Kwon 55eed6b7a5 [Model Runner V2] Add WhisperModelState [6/N] (#35790)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-03-11 14:20:38 -07:00
Woosuk Kwon 8ccbcda5c0 [Model Runner V2] Remove unused warmup_for_prefill method (#36762)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-03-11 08:02:44 -07:00
Woosuk Kwon 195d1ca3e8 [Minor] Enhance error message for TRTLLM decode uniformity check (#36609)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-03-10 15:38:45 -07:00
Woosuk Kwon f088a831dd [Model Runner V2] Use unpadded num_tokens for PW CUDA graph attn metadata (#36626)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-03-10 09:30:56 -07:00
Woosuk Kwon 006aea17d7 [BugFix] Remove incorrect assert in split_decodes_and_prefills (#36553)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-03-09 20:02:02 -07:00
Woosuk Kwon 2a194ddd72 [Model Runner V2] Add model_state inputs to CUDA graph capture (#36544)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-03-09 15:14:51 -07:00
Woosuk Kwon 10a5f4d53d [Model Runner V2] Use NamedTuple for execute_model_state (#35930)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-03-09 11:17:34 -07:00
Woosuk Kwon 6e956d9eca [Model Runner V2] Add dummy profile_cudagraph_memory API (#36520)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-03-09 10:20:13 -07:00
Woosuk Kwon 467886a0c4 [Model Runner V2] Fix inputs_embeds=None bug for MM models (#35917)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-03-03 13:47:45 -08:00
Woosuk Kwon a0a5178ab4 [Model Runner V2] Use ModelState.prepare_attn() for cuda graph capture [5/N] (#35774)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-03-02 20:06:27 -08:00
Woosuk Kwon 4f85bae9d6 [Docs][Model Runner V2] Add Design Docs (#35819)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-03-02 19:58:14 -08:00
Woosuk Kwon 72f4d16262 [Model Runner V2] Use block table apis for capture inputs (#35671)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-03-01 10:31:13 -08:00
Woosuk Kwon da543d1abe [Model Runner V2] Minor refactoring for EncoderRunner (#35628)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-03-01 00:15:39 -08:00
Woosuk Kwon e3eb146f7a [Model Runner V2] Add ModelStateInterface [4/N] (#35621)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-02-28 13:19:45 -08:00
Woosuk Kwon 1a014a0a93 [Model Runner V2] Move MM encoder to Model States [3/N] (#35564)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-02-27 18:32:38 -08:00
Woosuk Kwon 86ac7bcf84 [Model Runner V2] Support pooling models (#35120)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-02-27 18:03:01 -08:00
Woosuk Kwon 3d66502e1b [Model Runner V2] Prepare attn metadata in ModelState [2/N] (#35383)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-02-26 11:47:02 -08:00
Woosuk Kwon c66aa48e99 [Model Runner V2] Add model states [1/N] (#35350)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-02-26 11:20:35 -08:00
Woosuk Kwon 13025e71e8 [Model Runner V2] Add coding style guide (#35325)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-02-25 20:42:40 -08:00
Woosuk Kwon c645e9a214 [Model Runner V2] Remove propose_draft method (#35070)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-02-22 18:27:12 -08:00
Woosuk Kwon 2cbf9656ce [Model Runner V2] Enable CUDA graph for Eagle3 (#35040)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-02-21 21:42:50 -08:00
Woosuk Kwon b71fbd06e2 [Model Runner V2] Support attention group (#35036)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-02-21 16:42:53 -08:00
Woosuk Kwon a4047d4ea9 [Model Runner V2] Support Eagle3 (no CUDA graph) (#35029)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-02-21 12:55:24 -08:00
Woosuk Kwon 5fcb0cdd68 [Model Runner V2] Use FP32 for Gumbel Noise (#34854)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-02-18 17:07:37 -08:00
Woosuk Kwon c878b43b64 [Model Runner V2] Remove unnecessary copies in PW CUDA graph capture (#34849)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-02-18 15:52:50 -08:00
Woosuk Kwon 95be2a7f22 [Model Runner V2] Minor simplification for DCP (#34786)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-02-18 11:04:53 -08:00
Woosuk Kwon be3af2d29e [Model Runner V2] Further simplification for PP (#34724)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-02-17 15:18:18 -08:00
Woosuk Kwon d00df624f3 [Model Runner V2] Minor refactoring for penalties (#34662)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-02-16 21:43:00 -08:00
Woosuk Kwon 9752da9d9c [Model Runner V2] Minor simplification for BadWordsState (#34669)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-02-16 21:27:24 -08:00
Woosuk Kwon 04925b2202 [Model Runner V2] Minor cleanup for PP (#34666)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-02-16 19:15:31 -08:00
Woosuk Kwon d74278fb67 [Model Runner V2] Fix unintended CPU-GPU sync in make_dummy (#34667)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-02-16 19:00:29 -08:00
Woosuk Kwon 9ca768c740 [Model Runner V2] Minor cleanup for Sampler (#34563)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-02-14 18:29:03 -08:00
Woosuk Kwon 0916e7960b [GDN] Use CPU tensors to build GDN metadata (#34498)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-02-13 01:24:45 -08:00
Woosuk Kwon a2443de5fa [Model Runner V2] Use pinned memory for write_contents (#34222)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-02-10 08:55:22 -08:00
Woosuk KwonandNick Hill 46ec6d71c7 [Model Runner V2] Use a different stream for grammar bitmask h2d copy (#33059)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
Co-authored-by: Nick Hill <nhill@redhat.com>
2026-01-27 16:37:43 -08:00
Woosuk KwonandNick Hill 6d86fde09c [Model Runner V2] Remove UvaBufferPool for cpu->gpu copy (#33055)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
Co-authored-by: Nick Hill <nhill@redhat.com>
2026-01-26 16:47:35 -08:00
Woosuk Kwon a9b53dd435 [Model Runner V2] Add LoRAState to consolidate lora logic (#33062)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-01-25 22:21:12 -08:00
Woosuk Kwon 2f0d3ba745 [Model Runner V2] Minor simplification for finish_requests (#33048)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
2026-01-25 18:35:02 -08:00
Woosuk KwonandWoosuk Kwon edf927bc9f [Model Runner V2] Fix slot_mapping after #25954 (#33046)
Signed-off-by: Woosuk Kwon <woosuk@inferact.ai>
Co-authored-by: Woosuk Kwon <woosuk@inferact.ai>
2026-01-25 18:29:49 -08:00
Woosuk Kwon 5e00b561cd [Model Runner V2] Do not error on attention backends (#32820)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2026-01-21 17:02:48 -08:00
Woosuk Kwon 408195ec59 [Model Runner V2] Refactor Prompt Logprobs (#32811)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2026-01-21 15:12:20 -08:00
Woosuk Kwon e1da249c93 [Model Runner V2] Minor refactor for compute_slot_mappings (#32794)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2026-01-21 10:24:35 -08:00
Woosuk Kwon 9ab4388cd3 [Model Runner V2] Support FLASHINFER_MLA backend (#32709)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2026-01-20 10:26:17 -08:00
Woosuk Kwon e9c83cdc51 [Model Runner V2] Skip kernel launch for penalties & logit_bias (#32634)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2026-01-19 22:20:19 -08:00
Woosuk Kwon 6c01ffb897 [Model Runner V2] Decouple temperature from penalties (#32629)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2026-01-19 19:13:24 -08:00
Woosuk Kwon 7b7cdce968 [Model Runner V2] Refactor get_cudagraph_and_dp_padding (#32625)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2026-01-19 18:25:02 -08:00
Woosuk Kwon 05dc4bfab6 [Model Runner V2] Initialized communication buffer for DP (#32624)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2026-01-19 17:27:06 -08:00
Woosuk Kwon 43fada5360 [Model Runner V2] Refactor dummy_run (#32533)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2026-01-19 14:50:59 -08:00
Woosuk Kwon 9a1f16da1e [Model Runner V2] Refactor update_states (#32562)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2026-01-18 17:32:42 -08:00
Woosuk Kwon bb1848cd62 [Model Runner V2] Support VLM (#32546)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2026-01-18 16:58:51 -08:00
Woosuk Kwon 963dc0b865 [Model Runner V2] Minor optimization for eagle input processing (#32535)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2026-01-17 21:55:17 -08:00
Woosuk Kwon 4147910f1e [Model Runner V2] Move mrope_positions buffer to MRopeState (#32532)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2026-01-17 20:09:48 -08:00
Woosuk Kwon 6218034dd7 [Model Runner V2] Support FlashInfer backend & Fix CUDA Graph bug [1/2] (#32348)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2026-01-15 08:59:23 -08:00
Woosuk Kwon 90c0836902 [Model Runner V2] Refactor Sampler (#32245)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2026-01-13 17:58:12 -08:00
Woosuk Kwon 0a7dd23754 [Model Runner V2] Add support for M-RoPE (#32143)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2026-01-12 13:37:43 -08:00
Woosuk Kwon dec28688c5 [Model Runner V2] Minor refactor for logit_bias (#32209)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2026-01-12 13:08:30 -08:00
Woosuk Kwon ca81811bfe [Model Runner V2] Support logit_bias, allowed_token_ids, min_tokens (#32163)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2026-01-12 11:31:10 -08:00
Woosuk Kwon 025a32f9ed [Model Runner V2] Remove async barrier (#32083)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2026-01-11 20:24:30 -08:00
Woosuk Kwon 19504ac07f [Model Runner V2] Skip building deprecated fields in attn metadata (#32132)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2026-01-11 14:31:04 -08:00
Woosuk Kwon 7508243249 [Model Runner V2] Simplify BlockTables with UVA (#31965)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2026-01-08 10:24:26 -08:00
Woosuk Kwon d471b2aff0 [Model Runner V2] Support num NaNs in logits (#30187)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2025-12-09 10:00:49 -08:00
Woosuk Kwon 9e6562a3f6 [Model Runner V2] Fix Triton warning on tl.where (#30355)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2025-12-09 09:59:54 -08:00
Woosuk Kwon a238cbd89d [Model Runner V2] Support min-p sampling (#30171)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2025-12-05 21:42:47 -08:00
Woosuk Kwon cc050558f4 [Model Runner V2] Implement get_num_sampled_and_rejected kernel (#30029)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2025-12-04 07:19:42 -08:00
Woosuk Kwon ec38a7368d [Model Runner V2] Use packed mask for prompt bin counts (#29756)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2025-11-30 14:15:42 -08:00
Woosuk Kwon f223ed4181 [Model Runner V2] Fuse penalties and temperature into single kernel (#29720)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2025-11-29 02:29:16 -08:00
Woosuk Kwon 6afc0ffaf6 [Model Runner V2] Add sample/ directory and reorganize files (#29719)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2025-11-29 00:41:01 -08:00
Woosuk Kwon 4a80ad0a25 [Model Runner V2] Don't use UVA buffer for prefill_len (#29713)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2025-11-28 20:27:16 -08:00
Woosuk Kwon ca1b1e7296 [Model Runner V2] Refactor prefill token preparation (#29712)
Signed-off-by: Woosuk Kwon <woosuk.kwon@berkeley.edu>
2025-11-28 19:49:17 -08:00