From 7e751153eb5926882f2c32372aaf64d15986727e Mon Sep 17 00:00:00 2001 From: Cheng Wan <54331508+ch-wan@users.noreply.github.com> Date: Sun, 30 Aug 2026 02:18:33 -0700 Subject: [PATCH] [Config] Round 5.1: the published-side readers ask the bags, and a platform fact gets one address (#37086) --- .../sglang/srt/arg_groups/attention_hook.py | 30 +- .../sglang/srt/arg_groups/cuda_graph_hook.py | 13 +- .../sglang/srt/arg_groups/deepseek_v4_hook.py | 4 +- python/sglang/srt/arg_groups/dllm_hook.py | 4 +- python/sglang/srt/arg_groups/hisparse_hook.py | 14 +- python/sglang/srt/arg_groups/kimi_k3_hook.py | 7 +- python/sglang/srt/arg_groups/kv_cache_hook.py | 18 +- python/sglang/srt/arg_groups/mamba_hook.py | 24 +- python/sglang/srt/arg_groups/model_hook.py | 33 +- python/sglang/srt/arg_groups/moe_hook.py | 5 +- python/sglang/srt/arg_groups/overrides.py | 216 ++++++------ python/sglang/srt/arg_groups/parallel_hook.py | 5 +- python/sglang/srt/arg_groups/platform_hook.py | 11 +- python/sglang/srt/arg_groups/serving_hook.py | 25 +- .../sglang/srt/arg_groups/speculative_hook.py | 7 +- .../sglang/srt/arg_groups/validation_hook.py | 9 +- python/sglang/srt/configs/inkling.py | 6 +- python/sglang/srt/configs/model_config.py | 5 +- python/sglang/srt/debug_utils/dumper.py | 9 +- .../disaggregation/encoder/preprocessor.py | 6 +- .../srt/disaggregation/encoder/receiver.py | 10 +- .../srt/disaggregation/mooncake/conn.py | 8 +- python/sglang/srt/distributed/bootstrap.py | 25 +- .../triton_symm_mem_ag.py | 8 +- .../sglang/srt/distributed/parallel_state.py | 1 - python/sglang/srt/distributed/utils.py | 4 +- python/sglang/srt/dllm/mixin/scheduler.py | 2 +- python/sglang/srt/entrypoints/engine.py | 2 +- python/sglang/srt/entrypoints/grpc_bridge.py | 5 +- python/sglang/srt/entrypoints/http_server.py | 7 +- python/sglang/srt/entrypoints/sidecar.py | 6 +- python/sglang/srt/eplb/expert_location.py | 11 +- python/sglang/srt/eplb/lplb_solver.py | 5 +- .../npu/attention/ascend_backend.py | 8 +- .../layers/attention/attention_registry.py | 10 +- .../layers/attention/deepseek_v4_backend.py | 17 +- .../attention/dsa/paged_mqa_logits_backend.py | 5 +- .../srt/layers/attention/dsa_backend.py | 8 +- .../srt/layers/attention/dsv4/indexer.py | 9 +- .../layers/attention/flashinfer_backend.py | 6 +- .../attention/flashinfer_mla_backend.py | 6 +- .../srt/layers/attention/minicpm/backend.py | 10 +- .../attention/minimax_sparse_backend.py | 6 +- .../attention/tokenspeed_mla_backend.py | 7 +- .../layers/attention/trtllm_mha_backend.py | 13 +- python/sglang/srt/layers/attention/vision.py | 11 +- python/sglang/srt/layers/communicator.py | 14 +- python/sglang/srt/layers/cp/bcg.py | 2 +- .../layers/deep_gemm_wrapper/compile_utils.py | 6 +- .../layers/deep_gemm_wrapper/configurer.py | 4 +- .../layers/deep_gemm_wrapper/entrypoint.py | 5 +- python/sglang/srt/layers/dp_attention.py | 13 +- .../srt/layers/flashinfer_comm_fusion.py | 21 +- .../srt/layers/flashinfer_mnnvl_cutedsl.py | 3 +- python/sglang/srt/layers/k3_ar_fusion.py | 2 +- python/sglang/srt/layers/k3_gemm_ar.py | 3 +- python/sglang/srt/layers/k3_sp_collective.py | 2 +- python/sglang/srt/layers/moe/cutlass_moe.py | 9 +- .../fused_moe_triton/triton_kernels_moe.py | 4 +- .../sglang/srt/layers/moe/moe_runner/aiter.py | 2 +- .../sglang/srt/layers/moe/moe_runner/base.py | 2 +- .../srt/layers/moe/moe_runner/deep_gemm.py | 7 +- .../srt/layers/moe/token_dispatcher/deepep.py | 4 +- .../layers/moe/token_dispatcher/deepep_v2.py | 7 +- .../layers/moe/token_dispatcher/mooncake.py | 3 +- .../srt/layers/moe/token_dispatcher/nixl.py | 7 +- python/sglang/srt/layers/moe/topk.py | 8 +- python/sglang/srt/layers/moe/utils.py | 4 +- .../compressed_tensors/compressed_tensors.py | 5 +- .../compressed_tensors_w4a4_nvfp4_moe.py | 4 +- .../quantization/fp4_kv_cache_quant_method.py | 4 +- .../srt/layers/quantization/fp4_utils.py | 10 +- python/sglang/srt/layers/quantization/fp8.py | 21 +- .../srt/layers/quantization/fp8_utils.py | 43 ++- .../srt/layers/quantization/kvfp4_tensor.py | 20 +- .../srt/layers/quantization/modelopt_quant.py | 9 +- .../sglang/srt/layers/quantization/mxfp4.py | 26 +- .../mxfp4_flashinfer_cutlass_moe.py | 4 +- .../mxfp4_flashinfer_trtllm_moe.py | 9 +- .../layers/quantization/mxfp4_marlin_moe.py | 4 +- .../sglang/srt/layers/quantization/unquant.py | 7 +- python/sglang/srt/managers/rust_server.py | 52 +-- python/sglang/srt/managers/schedule_batch.py | 2 +- python/sglang/srt/managers/schedule_policy.py | 10 +- python/sglang/srt/managers/scheduler.py | 5 +- .../scheduler_components/load_publisher.py | 2 +- .../scheduler_components/metrics_reporter.py | 4 +- python/sglang/srt/mem_cache/hiradix_cache.py | 2 +- .../srt/mem_cache/kv_cache_configurator.py | 8 +- .../srt/mem_cache/mamba_checkpoint_pool.py | 2 +- .../sglang/srt/mem_cache/mamba_radix_cache.py | 3 +- .../sglang/srt/mem_cache/pool_host/common.py | 2 +- .../sglang/srt/mem_cache/sparsity/factory.py | 14 +- .../srt/mem_cache/unified_radix_cache.py | 8 +- .../srt/model_executor/cuda_graph_config.py | 3 +- .../sglang/srt/model_executor/model_runner.py | 6 +- .../load_model_utils.py | 2 +- .../model_runner_components/misc_utils.py | 8 +- .../srt/model_executor/pool_configurator.py | 6 +- .../runner_utils/capture_mode.py | 2 +- python/sglang/srt/model_loader/loader.py | 2 - .../sglang/srt/models/bailing_moe_linear.py | 11 +- python/sglang/srt/models/bailing_moe_v3.py | 4 +- .../amd/deepseek_v4_fused_mhc.py | 4 +- .../attention_backend_handler.py | 9 +- python/sglang/srt/models/deepseek_v2.py | 4 +- python/sglang/srt/models/deepseek_v4.py | 12 +- .../sglang/srt/models/deepseek_v4_dspark.py | 9 +- python/sglang/srt/models/gemma4_vision.py | 9 +- python/sglang/srt/models/gpt_oss.py | 5 +- .../inkling_common/quantization/config.py | 2 +- python/sglang/srt/models/kimi_k3.py | 7 +- python/sglang/srt/models/longcat_flash.py | 7 +- python/sglang/srt/models/mimo_audio.py | 6 +- python/sglang/srt/models/qwen2_moe.py | 8 +- .../multimodal/processors/base_processor.py | 7 +- .../srt/multimodal/transport/memory_pool.py | 3 +- .../srt/observability/metrics_collector.py | 2 +- python/sglang/srt/observability/trace.py | 3 +- python/sglang/srt/runtime_context.py | 308 +++++++++++++++++- python/sglang/srt/server_args.py | 137 +------- python/sglang/srt/speculative/dflash_utils.py | 3 +- .../sglang/srt/state_capturer/indexer_topk.py | 9 +- .../srt/state_capturer/routed_experts.py | 3 +- python/sglang/srt/utils/common.py | 11 +- .../srt/utils/cuda_vmm_transport_utils.py | 6 +- python/sglang/srt/weight_cache/ipc_loader.py | 2 +- .../test_multimodal_piecewise_cuda_graph.py | 35 +- .../constrained/test_base_grammar_backend.py | 6 +- .../attention/test_kda_helion_dispatcher.py | 15 +- ..._compressed_tensors_wna16_moe_no_linear.py | 17 +- .../quantization/test_mxfp4_sm120_cutlass.py | 7 +- .../layers/test_flashinfer_comm_fusion.py | 22 +- .../layers/test_minicpm_sparse_metadata.py | 11 +- .../unit/managers/test_mm_process_config.py | 8 +- .../unit/managers/test_prefill_adder.py | 63 ++-- .../managers/test_scheduler_hicache_attach.py | 8 +- .../models/test_deepseek_v4_amd_fused_mhc.py | 16 +- .../test_resolution_reads_no_bag.py | 7 +- .../test_resolution_reads_the_declarations.py | 2 +- .../unit/server_args/test_server_args.py | 167 +++++----- .../spec/test_dflash_extra_buffer_lazy.py | 21 +- test/registered/unit/test_model_overrides.py | 248 ++++++-------- .../unit/test_platform_address_not_frozen.py | 81 +++++ test/registered/unit/test_platform_context.py | 90 +++++ .../unit/test_pre_publish_readers.py | 134 ++++++++ test/registered/unit/test_runtime_context.py | 39 ++- ..._server_args_no_instance_mutation_entry.py | 9 +- 148 files changed, 1585 insertions(+), 1114 deletions(-) create mode 100644 test/registered/unit/test_platform_address_not_frozen.py create mode 100644 test/registered/unit/test_platform_context.py create mode 100644 test/registered/unit/test_pre_publish_readers.py diff --git a/python/sglang/srt/arg_groups/attention_hook.py b/python/sglang/srt/arg_groups/attention_hook.py index 120724a14..af9fd0054 100644 --- a/python/sglang/srt/arg_groups/attention_hook.py +++ b/python/sglang/srt/arg_groups/attention_hook.py @@ -20,6 +20,7 @@ from sglang.srt.arg_groups.overrides import ( _intel_xpu_page_constraint, _mla_backend_page_constraints, _mla_kv_cache_dtype_checks, + attention_backends_of, declare_resolution, mamba_extra_buffer_of, model_config_of, @@ -31,13 +32,8 @@ from sglang.srt.arg_groups.overrides import ( from sglang.srt.connector import ConnectorType from sglang.srt.environ import envs from sglang.srt.model_executor.cuda_graph_config import Backend, Phase, with_phase +from sglang.srt.runtime_context import get_platform from sglang.srt.utils.common import ( - is_cuda, - is_hip, - is_sm90_supported, - is_sm100_or_sm110_supported, - is_sm100_supported, - is_sm120_supported, parse_connector_type, ) @@ -45,7 +41,6 @@ logger = logging.getLogger(__name__) def handle_attention_backend_compatibility(server_args: Any): - from sglang.srt.arg_groups.overrides import attention_backends_of cfg = resolving_view(server_args) model_config = model_config_of(server_args) @@ -136,7 +131,7 @@ def handle_attention_backend_compatibility(server_args: Any): prefill_backend, decode_backend = attention_backends_of(resolved_view(server_args)) if "trtllm_mha" in (prefill_backend, decode_backend): if prefill_backend == "trtllm_mha" and not ( - is_sm90_supported() or is_sm100_supported() or is_sm120_supported() + get_platform().is_sm90 or get_platform().is_sm100 or get_platform().is_sm120 ): raise ValueError( "TRTLLM MHA backend for prefill requires Hopper (SM90), Blackwell (SM100), or SM120 GPUs. " @@ -144,7 +139,7 @@ def handle_attention_backend_compatibility(server_args: Any): ) if ( prefill_backend == "trtllm_mha" - and is_sm120_supported() + and get_platform().is_sm120 and ( cfg.kv_cache_dtype == "fp8_e4m3" or ( @@ -158,14 +153,14 @@ def handle_attention_backend_compatibility(server_args: Any): "fp8_e4m3 KV cache or skip-softmax." ) if decode_backend == "trtllm_mha" and not ( - is_sm90_supported() or is_sm100_supported() or is_sm120_supported() + get_platform().is_sm90 or get_platform().is_sm100 or get_platform().is_sm120 ): raise ValueError( "TRTLLM MHA backend for decode is only supported on Hopper (SM90), Blackwell (SM100) and (SM120) GPUs. Please use a different decode backend." ) if ( prefill_backend == "trtllm_mha" - and not is_sm100_supported() + and not get_platform().is_sm100 and (cfg.enable_prefill_context_parallel or cfg.attn_cp_size > 1) ): raise ValueError( @@ -227,7 +222,7 @@ def handle_linear_attn_backend(server_args: Any): if ( cfg.linear_attn_decode_backend is None and cfg.linear_attn_backend != "helion" - and is_sm100_supported() + and get_platform().is_sm100 and cfg.mamba_ssm_dtype == "bfloat16" # Stage 4: flashinfer's recurrent_kda compiles the state slot stride # as a free int64, so it reads the page-major/unified envelope-strided @@ -273,7 +268,7 @@ def handle_linear_attn_backend(server_args: Any): if ( decode == "flashinfer" and cfg.mamba_ssm_dtype != "bfloat16" - and is_cuda() + and get_platform().is_cuda and torch.cuda.get_device_capability()[0] >= 10 ): raise ValueError( @@ -288,7 +283,7 @@ def handle_linear_attn_backend(server_args: Any): if ( verify == "flashinfer" and cfg.mamba_ssm_dtype != "bfloat16" - and is_cuda() + and get_platform().is_cuda and torch.cuda.get_device_capability()[0] >= 10 ): raise ValueError( @@ -304,7 +299,7 @@ def handle_linear_attn_backend(server_args: Any): cuda_major = int(cuda_version.split(".")[0]) if cuda_version is not None else 0 if ( prefill == "flashinfer" - and is_cuda() + and get_platform().is_cuda and torch.cuda.get_device_capability()[0] >= 10 and cuda_major < 13 ): @@ -454,7 +449,6 @@ def handle_multi_item_scoring(server_args: Any): changing it silently could surprise users who intentionally picked a non-flashinfer backend. """ - from sglang.srt.arg_groups.overrides import attention_backends_of cfg = resolving_view(server_args) if not cfg.enable_mis: @@ -568,7 +562,7 @@ def handle_deterministic_inference(server_args: Any): raise ValueError( f"Currently only {RADIX_SUPPORTED_DETERMINISTIC_ATTENTION_BACKEND} attention backends are supported for deterministic inference with absorbed-MLA models. But you're using {attention_backend}." ) - if attention_backend == "fa4" and not is_sm100_or_sm110_supported(): + if attention_backend == "fa4" and not get_platform().is_sm100_or_sm110: raise ValueError( "Deterministic inference with absorbed-MLA models on the fa4 " "attention backend requires SM100/SM110: it runs " @@ -589,7 +583,7 @@ def handle_deterministic_inference(server_args: Any): # Check TP size if cfg.tp_size > 1: - if is_hip(): + if get_platform().is_hip: # AMD: use 1-stage all-reduce kernel which is inherently deterministic # (each GPU reads all data from all GPUs, reduces locally in fixed order) logger.info("AMD/ROCm: Using 1-stage all-reduce kernel (deterministic)") diff --git a/python/sglang/srt/arg_groups/cuda_graph_hook.py b/python/sglang/srt/arg_groups/cuda_graph_hook.py index 9a958188d..29f087773 100644 --- a/python/sglang/srt/arg_groups/cuda_graph_hook.py +++ b/python/sglang/srt/arg_groups/cuda_graph_hook.py @@ -7,6 +7,7 @@ import logging from typing import Any from sglang.srt.arg_groups.overrides import ( + attention_backends_of, declare_resolution, model_config_of, resolved_view, @@ -22,12 +23,10 @@ from sglang.srt.model_executor.cuda_graph_config import ( with_phase, ) from sglang.srt.platforms import current_platform +from sglang.srt.runtime_context import get_platform from sglang.srt.utils.common import ( is_cpu, - is_hip, is_mps, - is_npu, - is_xpu, parse_connector_type, ) from sglang.srt.utils.hf_transformers_utils import check_gguf_file @@ -111,7 +110,6 @@ def apply_cuda_graph_compatibility(server_args: Any): prefill backend (this folds in the old --enforce-piecewise-cuda-graph contract). """ - from sglang.srt.arg_groups.overrides import attention_backends_of cfg = resolving_view(server_args) if (Phase.PREFILL, "backend") in server_args._cuda_graph_config_locked: @@ -167,7 +165,11 @@ def disable_tc_piecewise_cudagraph_if_incompatible(server_args: Any): ("pipeline parallelism (pp_size > 1)", lambda: cfg.pp_size > 1), ( "non-CUDA hardware (HIP/NPU/CPU/MPS/XPU)", - lambda: is_hip() or is_npu() or is_cpu() or is_mps() or is_xpu(), + lambda: get_platform().is_hip + or get_platform().is_npu + or is_cpu() + or is_mps() + or get_platform().is_xpu, ), ( "OOT platform without piecewise support", @@ -330,7 +332,6 @@ def disable_prefill_cuda_graph_for_deepseek_trtllm_mla(server_args: Any): breakable) trtllm_mla falls back to FlashAttention for prefill and regresses performance, so disable whichever prefill graph backend is in effect. """ - from sglang.srt.arg_groups.overrides import attention_backends_of cfg = resolving_view(server_args) diff --git a/python/sglang/srt/arg_groups/deepseek_v4_hook.py b/python/sglang/srt/arg_groups/deepseek_v4_hook.py index ca96a1f02..2c991845f 100644 --- a/python/sglang/srt/arg_groups/deepseek_v4_hook.py +++ b/python/sglang/srt/arg_groups/deepseek_v4_hook.py @@ -10,6 +10,7 @@ from sglang.srt.arg_groups.overrides import ( run_post_process_pass, ) from sglang.srt.environ import envs +from sglang.srt.runtime_context import get_platform if TYPE_CHECKING: from sglang.srt.server_args import ServerArgs @@ -114,14 +115,13 @@ def apply_deepseek_v4_defaults(server_args: ServerArgs, model_arch: str) -> None that field) and the validations. """ cfg = resolving_view(server_args) - from sglang.srt.utils import is_hip # FlashMLA sparse prefill (SGLANG_OPT_FLASHMLA_SPARSE_PREFILL, default on) # currently returns incorrect output for DeepSeek-V4-Flash on ROCm/HIP # (MI355X), which breaks the disaggregation nightly. Keep the previous # (dense prefill) behavior on ROCm until the sparse kernel is validated # there; - if is_hip(): + if get_platform().is_hip: logger.warning( "Disabling SGLANG_OPT_FLASHMLA_SPARSE_PREFILL by default on ROCm/HIP " f"for {model_arch}; set it explicitly to override." diff --git a/python/sglang/srt/arg_groups/dllm_hook.py b/python/sglang/srt/arg_groups/dllm_hook.py index bc3b631d4..247395090 100644 --- a/python/sglang/srt/arg_groups/dllm_hook.py +++ b/python/sglang/srt/arg_groups/dllm_hook.py @@ -15,7 +15,7 @@ from sglang.srt.arg_groups.overrides import ( run_post_process_pass, ) from sglang.srt.model_executor.cuda_graph_config import Backend, Phase, with_phase -from sglang.srt.utils.common import is_hip +from sglang.srt.runtime_context import get_platform logger = logging.getLogger(__name__) @@ -27,7 +27,7 @@ def handle_dllm_inference(server_args: Any): # On AMD/HIP, disable cuda graph for DLLM (the attention_backend # resolution moved to the pipeline: arg_groups/overrides.py # _dllm_attention_backend, invoked below at its legacy slot). - if is_hip(): + if get_platform().is_hip: if ( cfg.cuda_graph_config.decode.backend != Backend.DISABLED or cfg.cuda_graph_config.prefill.backend != Backend.DISABLED diff --git a/python/sglang/srt/arg_groups/hisparse_hook.py b/python/sglang/srt/arg_groups/hisparse_hook.py index e94a0510c..3b38b1e40 100644 --- a/python/sglang/srt/arg_groups/hisparse_hook.py +++ b/python/sglang/srt/arg_groups/hisparse_hook.py @@ -8,6 +8,7 @@ from sglang.srt.arg_groups.overrides import ( resolved_view, resolving_view, ) +from sglang.srt.runtime_context import get_platform if TYPE_CHECKING: from sglang.srt.server_args import ServerArgs @@ -22,21 +23,14 @@ HISPARSE_ROCM_DSA_BACKENDS = {"tilelang", "aiter"} HISPARSE_KV_CACHE_DTYPES = ("bfloat16", "fp8_e4m3") -def _is_hip() -> bool: - """The one place this family asks about ROCm, and the seam the tests patch.""" - from sglang.srt.utils.common import is_hip - - return is_hip() - - def _hisparse_default_backend(kv_cache_dtype: str) -> str: - if _is_hip(): + if get_platform().is_hip: return "tilelang" return "flashmla_kv" if kv_cache_dtype == "fp8_e4m3" else "flashmla_sparse" def _hisparse_allowed_backends(kv_cache_dtype: str) -> set[str]: - if _is_hip(): + if get_platform().is_hip: return HISPARSE_ROCM_DSA_BACKENDS return HISPARSE_CUDA_DSA_BACKENDS_BY_DTYPE.get( kv_cache_dtype, {"flashmla_sparse", "flashmla_kv", "flashinfer_sparse_mla"} @@ -96,7 +90,7 @@ def validate_hisparse(server_args: ServerArgs) -> None: hf_config = model_config_of(server_args).hf_config is_v4_hisparse = is_deepseek_v4(hf_config) - is_hip = _is_hip() + is_hip = get_platform().is_hip assert is_deepseek_dsa(hf_config) or is_v4_hisparse, ( "--enable-hisparse is only supported for DSA (DeepSeek Sparse Attention) " "models (e.g., DeepSeek V3.2, GLM-5) and DeepSeek V4 now. " diff --git a/python/sglang/srt/arg_groups/kimi_k3_hook.py b/python/sglang/srt/arg_groups/kimi_k3_hook.py index 1b86f4427..2fd6d9c7a 100644 --- a/python/sglang/srt/arg_groups/kimi_k3_hook.py +++ b/python/sglang/srt/arg_groups/kimi_k3_hook.py @@ -7,6 +7,7 @@ from sglang.srt.arg_groups.overrides import ( declare_resolution, resolving_view, ) +from sglang.srt.runtime_context import get_platform if TYPE_CHECKING: from sglang.srt.server_args import ServerArgs @@ -17,7 +18,6 @@ logger = logging.getLogger(__name__) def apply_kimi_k3_spec_backend_defaults(server_args: ServerArgs) -> None: """Apply speculative backend defaults for Kimi hybrid models.""" cfg = resolving_view(server_args) - from sglang.srt.utils import is_sm100_supported if cfg.speculative_algorithm is None: return @@ -42,7 +42,7 @@ def apply_kimi_k3_spec_backend_defaults(server_args: ServerArgs) -> None: if ( cfg.speculative_algorithm == "DSPARK" and cfg.speculative_draft_attention_backend is None - and is_sm100_supported() + and get_platform().is_sm100 ): declare_resolution( server_args, @@ -58,7 +58,6 @@ def apply_kimi_k3_spec_backend_defaults(server_args: ServerArgs) -> None: def apply_kimi_k3_linear_attn_defaults(server_args: ServerArgs) -> None: """KDA decode-fallback default for Kimi hybrid models (spec-independent).""" cfg = resolving_view(server_args) - from sglang.srt.utils import is_sm100_supported # Preempts the generic SM100+bf16 flashinfer switch (a GDN default): on # KDA shapes the triton packed decode measures ~35% faster than @@ -66,7 +65,7 @@ def apply_kimi_k3_linear_attn_defaults(server_args: ServerArgs) -> None: if ( cfg.linear_attn_decode_backend is None and cfg.mamba_ssm_dtype == "bfloat16" - and is_sm100_supported() + and get_platform().is_sm100 ): declare_resolution( server_args, diff --git a/python/sglang/srt/arg_groups/kv_cache_hook.py b/python/sglang/srt/arg_groups/kv_cache_hook.py index 1a58fbe87..def9f3dd9 100644 --- a/python/sglang/srt/arg_groups/kv_cache_hook.py +++ b/python/sglang/srt/arg_groups/kv_cache_hook.py @@ -7,6 +7,7 @@ import logging from typing import Any from sglang.srt.arg_groups.overrides import ( + attention_backends_of, declare_resolution, resolved_view, resolving_view, @@ -14,12 +15,7 @@ from sglang.srt.arg_groups.overrides import ( ) from sglang.srt.environ import envs from sglang.srt.model_executor.cuda_graph_config import Backend -from sglang.srt.utils.common import ( - is_blackwell_supported, - is_cuda, - is_sm100_supported, - is_sm120_supported, -) +from sglang.srt.runtime_context import get_platform logger = logging.getLogger(__name__) @@ -29,7 +25,7 @@ def handle_mxfp8_kv_cache_compatibility(server_args: Any) -> None: cfg = resolving_view(server_args) if cfg.kv_cache_dtype != "mxfp8": return - if not is_blackwell_supported(): + if not get_platform().is_blackwell: raise ValueError( "--kv-cache-dtype mxfp8 requires an SM100+ (Blackwell) GPU for the " "block-scaled operands used by the FA4 MXFP8 attention path." @@ -38,7 +34,6 @@ def handle_mxfp8_kv_cache_compatibility(server_args: Any) -> None: def handle_kv4_compatibility(server_args: Any) -> None: """Check FP4 KV cache compatibility with the attention backend""" - from sglang.srt.arg_groups.overrides import attention_backends_of cfg = resolving_view(server_args) @@ -49,9 +44,9 @@ def handle_kv4_compatibility(server_args: Any) -> None: prefill_backend, decode_backend = attention_backends_of(resolved_view(server_args)) attention_backend = resolved_view(server_args).attention_backend - if is_cuda(): + if get_platform().is_cuda: if cfg.kv_cache_dtype == "nvfp4" and not ( - is_sm100_supported() or is_sm120_supported() + get_platform().is_sm100 or get_platform().is_sm120 ): raise RuntimeError( "--kv-cache-dtype=nvfp4 requires Blackwell SM100 or SM120. " @@ -123,7 +118,6 @@ def handle_prefill_only_disable_kv_cache(server_args: Any) -> None: still None, backends haven't settled yet and the resolved (prefill, decode) pair would be a stale (None, None). """ - from sglang.srt.arg_groups.overrides import attention_backends_of cfg = resolving_view(server_args) @@ -199,7 +193,6 @@ def handle_cache_compatibility(server_args: Any) -> None: def handle_unified_memory_pool(server_args: Any) -> None: - from sglang.srt.arg_groups.overrides import attention_backends_of cfg = resolving_view(server_args) if not cfg.enable_unified_memory: @@ -280,7 +273,6 @@ def handle_page_major_kv_layout(server_args: Any): # The unified pool stores state in the page-major envelope-strided layout, so # enabling it implies --enable-page-major-kv-layout — routing it through the # single page-major path + stride-aware Triton asserts (set before the guard). - from sglang.srt.arg_groups.overrides import attention_backends_of cfg = resolving_view(server_args) if cfg.enable_unified_memory: diff --git a/python/sglang/srt/arg_groups/mamba_hook.py b/python/sglang/srt/arg_groups/mamba_hook.py index 3bb704dde..39e28cc9e 100644 --- a/python/sglang/srt/arg_groups/mamba_hook.py +++ b/python/sglang/srt/arg_groups/mamba_hook.py @@ -8,16 +8,9 @@ from typing import Any from sglang.srt.arg_groups.overrides import ( resolving_view, + supports_mamba_cache_extra_buffer, ) -from sglang.srt.utils.common import ( - is_cuda, - is_flashinfer_available, - is_hip, - is_musa, - is_npu, - is_sm100_supported, - is_xpu, -) +from sglang.srt.runtime_context import get_platform logger = logging.getLogger(__name__) @@ -41,13 +34,13 @@ def handle_mamba_backend(server_args: Any): "Run with --mamba-ssm-dtype float16 or disable " "--enable-mamba-cache-stochastic-rounding." ) - if not is_cuda(): + if not get_platform().is_cuda: raise ValueError( "Stochastic rounding for the Mamba SSM cache is only " "supported on NVIDIA CUDA platforms. Disable " "--enable-mamba-cache-stochastic-rounding on this platform." ) - if cfg.mamba_backend == "triton" and not is_sm100_supported(): + if cfg.mamba_backend == "triton" and not get_platform().is_sm100: raise ValueError( "Stochastic rounding for the Mamba SSM cache with " "--mamba-backend triton requires SM100 with CUDA >= 12.8 " @@ -67,7 +60,7 @@ def handle_mamba_backend(server_args: Any): " Stochastic rounding with --mamba-backend flashinfer " "requires FlashInfer Mamba and --mamba-ssm-dtype float16." ) - if is_flashinfer_available(): + if get_platform().has_flashinfer: try: import flashinfer.mamba # noqa: F401 @@ -103,13 +96,16 @@ def handle_int8_mamba_checkpoint(server_args: Any): def validate_mamba_extra_buffer(view, model_arch: str, *, mamba_cache_chunk_size_of): - from sglang.srt.arg_groups.overrides import supports_mamba_cache_extra_buffer assert supports_mamba_cache_extra_buffer( view, model_arch ), f"extra_buffer is not supported for {model_arch}; use no_buffer." assert ( - is_cuda() or is_musa() or is_npu() or is_hip() or is_xpu() + get_platform().is_cuda + or get_platform().is_musa + or get_platform().is_npu + or get_platform().is_hip + or get_platform().is_xpu ), "extra_buffer needs CUDA/MUSA/NPU/ROCm/XPU (FLA)." if view.mamba_radix_cache_strategy == "extra_buffer_lazy": # The PD-disagg decode pool is not wired for lazy slots. diff --git a/python/sglang/srt/arg_groups/model_hook.py b/python/sglang/srt/arg_groups/model_hook.py index 4db9df252..ad7eda7e4 100644 --- a/python/sglang/srt/arg_groups/model_hook.py +++ b/python/sglang/srt/arg_groups/model_hook.py @@ -16,6 +16,7 @@ from sglang.srt.arg_groups.overrides import ( _hrm_text_attention_force, _mamba_radix_cache_resolution, _sparse_head_overlap_disable, + attention_backends_of, collect_model_override_declarations, declare_resolution, mamba_cache_chunk_size, @@ -33,16 +34,10 @@ from sglang.srt.connector import ConnectorType from sglang.srt.environ import envs from sglang.srt.hardware_backend.mlx.runtime import use_mlx from sglang.srt.model_executor.cuda_graph_config import Backend, Phase, with_phase +from sglang.srt.runtime_context import get_platform from sglang.srt.utils.common import ( get_quantization_config, - is_cuda, - is_hip, is_mps, - is_npu, - is_sm90_supported, - is_sm100_supported, - is_sm120_supported, - is_xpu, parse_connector_type, ) @@ -50,7 +45,6 @@ logger = logging.getLogger(__name__) def handle_model_specific_adjustments(server_args: Any): - from sglang.srt.arg_groups.overrides import attention_backends_of cfg = resolving_view(server_args) from sglang.srt.configs.model_config import ( @@ -189,7 +183,9 @@ def handle_model_specific_adjustments(server_args: Any): "shared layers would run sparse attention without indices." ) - if not is_npu() and not is_xpu(): # CUDA or ROCm GPU + if ( + not get_platform().is_npu and not get_platform().is_xpu + ): # CUDA or ROCm GPU if cfg.enable_prefill_cp: # The DSA CP field declarations moved to the override # registry (arg_groups/overrides.py: @@ -302,7 +298,7 @@ def handle_model_specific_adjustments(server_args: Any): # latter awaiting the speculative-hook migration) stays below. run_post_process_pass(server_args, _deepseek_moe_quant_resolution) - if is_hip(): + if get_platform().is_hip: if is_deepseek_dsa(hf_config): # The fused top-k v2 kernel (topk_transform_512_v2) is a # CUDA/Hopper-only path: its JIT source includes @@ -336,7 +332,7 @@ def handle_model_specific_adjustments(server_args: Any): validate_deepseek_v4_cp(server_args) validate_deepseek_v4_mega_moe_token_budget(server_args) - if is_sm120_supported(): + if get_platform().is_sm120: # SM120 lacks tcgen05/TMEM: disable features that depend on # DeepGEMM or require >99KB SMEM (topk_v2). envs.SGLANG_OPT_FP8_WO_A_GEMM.set(False) @@ -348,7 +344,7 @@ def handle_model_specific_adjustments(server_args: Any): envs.SGLANG_FP8_PAGED_MQA_LOGITS_TORCH.set(True) # Prefer TileLang over the Torch fallback. envs.SGLANG_OPT_USE_TILELANG_INDEXER.set(True) - elif is_hip(): + elif get_platform().is_hip: envs.SGLANG_OPT_DEEPGEMM_HC_PRENORM.set(False) envs.SGLANG_OPT_FP8_WO_A_GEMM.set(False) envs.SGLANG_OPT_USE_JIT_INDEXER_METADATA.set(False) @@ -395,7 +391,7 @@ def handle_model_specific_adjustments(server_args: Any): if ( not resolved_view(server_args).enable_dp_attention and cfg.nnodes == 1 - and is_hip() + and get_platform().is_hip ): # TODO (Hubert): Put this back later # server_args.enable_aiter_allreduce_fusion = True @@ -714,8 +710,8 @@ def handle_model_capability_adjustments(server_args: Any): cfg.prefill_attention_backend or cfg.attention_backend ) if ( - is_cuda() - and (is_sm90_supported() or is_sm100_supported()) + get_platform().is_cuda + and (get_platform().is_sm90 or get_platform().is_sm100) and requested_prefill_backend in (None, "fa3", "fa4") ): # Hopper/Blackwell's default FA backend can consume raw K/V @@ -735,7 +731,10 @@ def handle_model_capability_adjustments(server_args: Any): cfg.cuda_graph_config, Phase.DECODE, backend=Backend.DISABLED ), ) - if is_cuda() and cfg.cuda_graph_config.prefill.backend != Backend.DISABLED: + if ( + get_platform().is_cuda + and cfg.cuda_graph_config.prefill.backend != Backend.DISABLED + ): declare_resolution( server_args, "_handle_model_capability_adjustments", @@ -776,7 +775,7 @@ def handle_model_capability_adjustments(server_args: Any): cfg.cuda_graph_config, Phase.PREFILL, **sizing ), ) - elif not is_cuda(): + elif not get_platform().is_cuda: # BCG is CUDA-only. Other graph backends do not support this # encoder-style prefill, so retain the eager Triton path. declare_resolution( diff --git a/python/sglang/srt/arg_groups/moe_hook.py b/python/sglang/srt/arg_groups/moe_hook.py index 15f48159c..cc45fbc7d 100644 --- a/python/sglang/srt/arg_groups/moe_hook.py +++ b/python/sglang/srt/arg_groups/moe_hook.py @@ -25,7 +25,8 @@ from sglang.srt.arg_groups.overrides import ( from sglang.srt.connector import ConnectorType from sglang.srt.environ import envs from sglang.srt.model_executor.cuda_graph_config import Backend, Phase, with_phase -from sglang.srt.utils.common import is_npu, parse_connector_type +from sglang.srt.runtime_context import get_platform +from sglang.srt.utils.common import parse_connector_type logger = logging.getLogger(__name__) @@ -243,7 +244,7 @@ def handle_a2a_moe(server_args: Any): # The resolving view, not the field: `_a2a_backend_overrides` may have # moved this already (waterfill forces `deepep`). a2a_now = resolved_view(server_args).moe_a2a_backend - if (a2a_now == "none" and is_npu()) or a2a_now == "ascend_tp": + if (a2a_now == "none" and get_platform().is_npu) or a2a_now == "ascend_tp": # FIXME (OrangeRedeng): for some reasons if pass "ascend_tp" accuracy drops to zero declare_resolution( server_args, diff --git a/python/sglang/srt/arg_groups/overrides.py b/python/sglang/srt/arg_groups/overrides.py index 87170be60..36a47cd60 100644 --- a/python/sglang/srt/arg_groups/overrides.py +++ b/python/sglang/srt/arg_groups/overrides.py @@ -46,30 +46,20 @@ from sglang.srt.environ import envs from sglang.srt.hardware_backend.mlx.runtime import use_mlx from sglang.srt.model_executor.cuda_graph_config import Backend from sglang.srt.platforms import current_platform +from sglang.srt.runtime_context import ( + get_context, + get_platform, +) from sglang.srt.utils.common import ( - cpu_has_amx_support, - get_device_capability, get_device_name, - get_device_sm, get_nvidia_driver_version, get_quantization_config, - is_blackwell_supported, is_cpu, - is_cuda, - is_flashinfer_available, is_gfx95_supported, - is_hip, - is_hopper_with_cuda_12_3, is_mnnvl_fabric_device, is_mps, - is_musa, is_no_spec_infer_or_topk_one, - is_npu, - is_sm90_supported, - is_sm100_supported, - is_sm120_supported, is_triton_kernels_available, - is_xpu, xpu_has_xmx_support, ) @@ -241,7 +231,6 @@ def run_post_process_pass(server_args: Any, fn: Callable[..., dict]) -> None: ``declare_late_resolution`` is -- post-publish changes go to the bags through ``get_context().override(...)``. """ - from sglang.srt.runtime_context import get_context declared = fn(ResolvedView(server_args, overlay=_declaration_overlay(server_args))) if not isinstance(declared, dict): @@ -280,17 +269,6 @@ def run_post_process_pass(server_args: Any, fn: Callable[..., dict]) -> None: validate_declarations(server_args, [entry]) -def _apply_fields(server_args: Any, fields: Dict[str, Any]) -> None: - """Write fields on behalf of the pipeline (bypasses the strict bare- - assignment guard that protects post-resolution mutation).""" - object.__setattr__(server_args, "_internal_write", True) - try: - for field, value in fields.items(): - setattr(server_args, field, value) - finally: - object.__setattr__(server_args, "_internal_write", False) - - def declare_resolution(server_args: Any, source: str, **fields: Any) -> None: """Record a resolution write in the declaration stash. @@ -331,7 +309,6 @@ def declare_late_resolution(server_args: Any, source: str, **fields: Any) -> Non field write would desync them, which is what ``get_context().override`` is for. """ - from sglang.srt.runtime_context import get_context try: published = get_context().server_args @@ -703,7 +680,7 @@ def _kimi_k3_overrides(server_args: Any, hf_config: Any) -> dict: overrides["dcp_comm_backend"] = dcp_comm_backend return overrides - if not (is_sm100_supported() and get_device_sm() in (100, 103)): + if not (get_platform().is_sm100 and get_platform().device_sm in (100, 103)): return {} backends_unset = is_attention_backend_not_set(cfg) if cfg.speculative_algorithm != "DSPARK": @@ -777,7 +754,7 @@ def _kimi_k3_moe_runner_overrides(server_args: Any, hf_config: Any) -> dict: cfg = resolving_view(server_args) if cfg.moe_runner_backend != "auto": return {} - if not (is_sm100_supported() and get_device_sm() in (100, 103, 107)): + if not (get_platform().is_sm100 and get_platform().device_sm in (100, 103, 107)): return {} if not _is_mxfp4_pack_quantized(hf_config): return {} @@ -815,7 +792,7 @@ def _deepseek_family_overrides(server_args: Any, hf_config: Any) -> dict: if is_attention_backend_not_set(cfg): overrides["attention_backend"] = "dsa" logger.info("Use dsa attention backend for DeepSeek with DSA.") - if not is_npu() and not is_xpu(): # CUDA or ROCm GPU + if not get_platform().is_npu and not get_platform().is_xpu: # CUDA or ROCm GPU if cfg.enable_prefill_cp: logger.warning( "Context parallel feature is still under experiment. It has only been verified on Hopper platform." @@ -858,7 +835,7 @@ def _deepseek_family_overrides(server_args: Any, hf_config: Any) -> dict: aiter_can_use_preshuffle_paged_mqa, ) - if is_hip() and not aiter_can_use_preshuffle_paged_mqa(): + if get_platform().is_hip and not aiter_can_use_preshuffle_paged_mqa(): # Legacy ROCm DSA path: aiter's gluon paged-MQA kernel is # unavailable (Triton<3.5 and AITER_ENABLE_AOT_GLUON_PA_MQA_LOGITS # not set, or SGLANG_DSA_HIP_DISABLE_PRESHUFFLE=1 / SGLANG_USE_AITER=0). @@ -873,7 +850,7 @@ def _deepseek_family_overrides(server_args: Any, hf_config: Any) -> dict: logger.warning("Setting page size to 64 for DeepSeek DSA.") else: # DeepSeek V3/R1/V3.1 - if is_sm100_supported(): + if get_platform().is_sm100: if ( cfg.attention_backend is None and cfg.prefill_attention_backend is None @@ -928,7 +905,7 @@ def _mimo_v2_overrides(server_args: Any, hf_config: Any) -> dict: # On Blackwell "auto" falls through to the triton fused-MoE runner, ~12% # slower at bs=1 decode. FP4 checkpoints use flashinfer_mxfp4 instead. if ( - is_sm100_supported() + get_platform().is_sm100 and cfg.moe_runner_backend == "auto" and get_quantization_config(hf_config) == "fp8" ): @@ -945,7 +922,7 @@ def _minimax_m2_overrides(server_args: Any, hf_config: Any) -> dict: "Enable TF32 matmul for MiniMaxM2ForCausalLM model to improve gate gemm performance." ) if ( - is_sm100_supported() + get_platform().is_sm100 and cfg.moe_runner_backend == "auto" and model_config_of(server_args).quantization == "modelopt_fp4" ): @@ -973,7 +950,7 @@ def _minimax_m3_overrides(server_args: Any, hf_config: Any) -> dict: overrides["quantization"] = quant_method quant_resolved = quant_method - if is_hip(): + if get_platform().is_hip: if is_attention_backend_not_set(cfg): overrides["attention_backend"] = "triton" if cfg.moe_runner_backend == "auto" and quant_resolved == "mxfp8": @@ -996,7 +973,7 @@ def _minimax_m3_overrides(server_args: Any, hf_config: Any) -> dict: # accelerate the large prefill all-reduce. if not aiter_fusion_resolved and not envs.SGLANG_M3_ALLOW_CUSTOM_AR.get(): overrides["disable_custom_all_reduce"] = True - elif is_sm100_supported(): + elif get_platform().is_sm100: if is_attention_backend_not_set(cfg): if ( cfg.kv_cache_dtype == "fp8_e4m3" @@ -1027,7 +1004,7 @@ def _minimax_m3_overrides(server_args: Any, hf_config: Any) -> dict: f"{overrides.get('attention_backend', cfg.attention_backend)}, page_size={page_resolved}, " f"moe_runner_backend={overrides.get('moe_runner_backend', cfg.moe_runner_backend)}." ) - elif is_sm90_supported(): + elif get_platform().is_sm90: if is_attention_backend_not_set(cfg): overrides["attention_backend"] = "fa3" page_resolved = cfg.page_size @@ -1060,7 +1037,7 @@ def _minimax_m3_overrides(server_args: Any, hf_config: Any) -> dict: elif ( cfg.kv_cache_dtype == "fp8_e4m3" and overrides.get("attention_backend", cfg.attention_backend) == "trtllm_mha" - and is_sm100_supported() + and get_platform().is_sm100 ): if envs.SGLANG_DISABLE_M3_FP8_ATTN_GEMM.get(): logger.info( @@ -1121,22 +1098,22 @@ def _gpt_oss_overrides(server_args: Any, hf_config: Any) -> dict: overrides: Dict[str, Any] = {} # Set attention backend for GPT-OSS if is_attention_backend_not_set(cfg): - if is_sm100_supported(): + if get_platform().is_sm100: overrides["attention_backend"] = "trtllm_mha" - elif is_sm90_supported(): + elif get_platform().is_sm90: overrides["attention_backend"] = "fa3" - elif is_cpu() and cpu_has_amx_support(): + elif is_cpu() and get_platform().has_amx: overrides["attention_backend"] = "intel_amx" - elif is_xpu(): + elif get_platform().is_xpu: overrides["attention_backend"] = "intel_xpu" - elif is_hip(): + elif get_platform().is_hip: overrides["attention_backend"] = "aiter" elif not (is_mps() and use_mlx()): # Exempt MLX only -- it owns attention in its own runner. macOS # without MLX still falls through to triton and fails fast below, # rather than landing on torch_native (no sliding window, no sinks). overrides["attention_backend"] = "triton" - if is_xpu(): + if get_platform().is_xpu: # Check for bf16 dtype on Intel XPU. Reads the pristine dtype request, # which equals the legacy mid-branch read: dtype had no earlier writer # for this arch. @@ -1159,18 +1136,20 @@ def _gpt_oss_overrides(server_args: Any, hf_config: Any) -> dict: overrides["dtype"] = "bfloat16" if cfg.moe_runner_backend == "auto": - if is_sm100_supported() and is_mxfp4_quant_format: + if get_platform().is_sm100 and is_mxfp4_quant_format: overrides["moe_runner_backend"] = "flashinfer_mxfp4" logger.warning( "Detected SM100 and MXFP4 quantization format for GPT-OSS model, enabling FlashInfer MXFP4 MOE kernel." ) - elif is_sm120_supported() and is_mxfp4_quant_format: + elif get_platform().is_sm120 and is_mxfp4_quant_format: overrides["moe_runner_backend"] = "flashinfer_mxfp4" logger.warning( "Detected SM120 and MXFP4 quantization format for GPT-OSS model, " "enabling FlashInfer CUTLASS MXFP4 MOE kernel." ) - elif (is_hip() and envs.SGLANG_USE_AITER.get()) and is_mxfp4_quant_format: + elif ( + get_platform().is_hip and envs.SGLANG_USE_AITER.get() + ) and is_mxfp4_quant_format: overrides["moe_runner_backend"] = "auto" logger.warning( "Detected ROCm and MXFP4 quantization format for GPT-OSS model, enabling aiter MXFP4 MOE kernel." @@ -1182,14 +1161,14 @@ def _gpt_oss_overrides(server_args: Any, hf_config: Any) -> dict: ## callers default to INTERLEAVE; opt this path out ## unless the user explicitly overrode it. # envs.SGLANG_USE_AITER_MOE_GU_ITLV.set(False) - elif is_hip() and envs.SGLANG_USE_AITER.get(): + elif get_platform().is_hip and envs.SGLANG_USE_AITER.get(): # For GPT-OSS bf16 on ROCm with aiter, use triton backend # because aiter CK kernel doesn't support all GEMM dimensions overrides["moe_runner_backend"] = "triton" logger.warning( "Detected ROCm with SGLANG_USE_AITER for GPT-OSS bf16 model, using triton MOE kernel." ) - elif is_musa() and envs.SGLANG_DEEPEP_BF16_DISPATCH.get(): + elif get_platform().is_musa and envs.SGLANG_DEEPEP_BF16_DISPATCH.get(): overrides["moe_runner_backend"] = "deep_gemm" logger.warning( "Detected MUSA with SGLANG_DEEPEP_BF16_DISPATCH for bf16 model, using deep_gemm kernel." @@ -1198,11 +1177,11 @@ def _gpt_oss_overrides(server_args: Any, hf_config: Any) -> dict: cfg.ep_size == 1 and is_triton_kernels_available() and cfg.quantization is None - and not (is_cpu() and cpu_has_amx_support()) + and not (is_cpu() and get_platform().has_amx) ): # The triton_kernels package segfaults on Blackwell (B200) # with NVIDIA driver >= 595. Fall back to triton backend. - if is_blackwell_supported() and get_nvidia_driver_version() >= (595,): + if get_platform().is_blackwell and get_nvidia_driver_version() >= (595,): overrides["moe_runner_backend"] = "triton" logger.warning( "Detected GPT-OSS model on Blackwell with driver >= 595, " @@ -1225,11 +1204,11 @@ def _llama4_overrides(server_args: Any, hf_config: Any) -> dict: overrides: Dict[str, Any] = {} # Auto-select attention backend for Llama4 if not specified if cfg.attention_backend is None: - if is_sm100_supported(): + if get_platform().is_sm100: backend, platform = "trtllm_mha", "sm100" - elif is_sm90_supported(): + elif get_platform().is_sm90: backend, platform = "fa3", "sm90" - elif is_hip(): + elif get_platform().is_hip: backend, platform = "aiter", "hip" elif cfg.device == "xpu": backend, platform = "intel_xpu", "xpu" @@ -1239,7 +1218,7 @@ def _llama4_overrides(server_args: Any, hf_config: Any) -> dict: f"Use {backend} as attention backend on {platform} for Llama4 model" ) overrides["attention_backend"] = backend - if is_sm100_supported() and cfg.moe_runner_backend == "auto": + if get_platform().is_sm100 and cfg.moe_runner_backend == "auto": if cfg.quantization in {"fp8", "modelopt_fp8"}: overrides["moe_runner_backend"] = "flashinfer_trtllm" logger.info( @@ -1256,7 +1235,7 @@ def _llama4_overrides(server_args: Any, hf_config: Any) -> dict: def _gemma4_overrides(server_args: Any, hf_config: Any) -> dict: cfg = resolving_view(server_args) overrides: Dict[str, Any] = {} - default_attention_backend = "trtllm_mha" if is_sm100_supported() else "triton" + default_attention_backend = "trtllm_mha" if get_platform().is_sm100 else "triton" if is_attention_backend_not_set(cfg): logger.info( f"Use {default_attention_backend} as default attention backend for Gemma4" @@ -1267,7 +1246,7 @@ def _gemma4_overrides(server_args: Any, hf_config: Any) -> dict: # choose an unsupported backend later. elif cfg.attention_backend is None: overrides["attention_backend"] = default_attention_backend - if is_sm100_supported() and cfg.moe_runner_backend == "auto": + if get_platform().is_sm100 and cfg.moe_runner_backend == "auto": if model_config_of(server_args).quantization == "modelopt_fp4": overrides["quantization"] = "modelopt_fp4" overrides["moe_runner_backend"] = "flashinfer_trtllm" @@ -1311,7 +1290,7 @@ def _minicpm_sala_overrides(server_args: Any, hf_config: Any) -> dict: overrides["disable_radix_cache"] = True if envs.SGLANG_MINICPM_FORCE_DENSE.get(): dense_backends = { - "minicpm_flashattn": ("fa4" if is_blackwell_supported() else "fa3"), + "minicpm_flashattn": ("fa4" if get_platform().is_blackwell else "fa3"), "minicpm_flashinfer": "flashinfer", } # Literal keys keep the written-field set statically derivable; a loop @@ -1341,7 +1320,7 @@ def _minicpm_sala_overrides(server_args: Any, hf_config: Any) -> dict: if is_attention_backend_not_set(cfg): overrides["attention_backend"] = ( "minicpm_flashinfer" - if is_blackwell_supported() + if get_platform().is_blackwell else "minicpm_flashattn" ) return overrides @@ -1350,7 +1329,7 @@ def _minicpm_sala_overrides(server_args: Any, hf_config: Any) -> dict: @_register_for("MiniCPMV4_6ForConditionalGeneration") def _minicpm_v4_6_overrides(server_args: Any, hf_config: Any) -> dict: cfg = resolving_view(server_args) - if is_sm100_supported() and cfg.attention_backend is None: + if get_platform().is_sm100 and cfg.attention_backend is None: return {"attention_backend": "triton"} return {} @@ -1360,7 +1339,7 @@ def _minicpm_v4_6_overrides(server_args: Any, hf_config: Any) -> dict: ) def _falcon_h1_jet_overrides(server_args: Any, hf_config: Any) -> dict: cfg = resolving_view(server_args) - if is_sm100_supported() and cfg.attention_backend is None: + if get_platform().is_sm100 and cfg.attention_backend is None: return {"attention_backend": "triton"} return {} @@ -1371,7 +1350,7 @@ def _granite_moe_hybrid_overrides(server_args: Any, hf_config: Any) -> dict: has_mamba = any( layer_type == "mamba" for layer_type in getattr(hf_config, "layer_types", []) ) - if has_mamba and is_sm100_supported() and cfg.attention_backend is None: + if has_mamba and get_platform().is_sm100 and cfg.attention_backend is None: return {"attention_backend": "flashinfer"} return {} @@ -1379,7 +1358,7 @@ def _granite_moe_hybrid_overrides(server_args: Any, hf_config: Any) -> dict: @_register_for("Lfm2ForCausalLM", "Lfm2MoeForCausalLM") def _lfm2_overrides(server_args: Any, hf_config: Any) -> dict: cfg = resolving_view(server_args) - if is_sm100_supported() and cfg.attention_backend is None: + if get_platform().is_sm100 and cfg.attention_backend is None: return {"attention_backend": "flashinfer"} return {} @@ -1426,11 +1405,15 @@ def _deepseek_v4_overrides(server_args: Any, hf_config: Any) -> dict: ) elif ( cfg.device == "cuda" - and not is_hip() + and not get_platform().is_hip and cfg.moe_a2a_backend == "none" and not envs.SGLANG_DSV4_FP4_DEQUANT.get() and model_config.is_fp4_experts - and (is_sm90_supported() or is_sm100_supported() or is_sm120_supported()) + and ( + get_platform().is_sm90 + or get_platform().is_sm100 + or get_platform().is_sm120 + ) ): overrides["moe_runner_backend"] = "flashinfer_mxfp4" logger.info( @@ -1481,7 +1464,7 @@ def _inkling_overrides(server_args: Any, hf_config: Any) -> dict: # (mirrors the MiniMax-M3 SM100 fa4-default above); an explicit # --attention-backend / --prefill/decode-attention-backend still wins. if is_attention_backend_not_set(cfg): - inkling_attn_backend = "fa4" if is_sm100_supported() else "triton" + inkling_attn_backend = "fa4" if get_platform().is_sm100 else "triton" overrides["attention_backend"] = inkling_attn_backend logger.info( f"Use {inkling_attn_backend} as the attention backend for Inkling " @@ -1547,7 +1530,7 @@ def _nemotron_h_overrides(server_args: Any, hf_config: Any) -> dict: elif (is_modelopt or model_config.quantization is None) and ( cfg.moe_runner_backend == "auto" ): - if is_sm100_supported() and cfg.moe_a2a_backend == "none": + if get_platform().is_sm100 and cfg.moe_a2a_backend == "none": overrides["moe_runner_backend"] = "flashinfer_trtllm" logger.info( f"Use flashinfer_trtllm as MoE runner backend on sm100 for {model_arch}" @@ -1557,8 +1540,8 @@ def _nemotron_h_overrides(server_args: Any, hf_config: Any) -> dict: model_config.quantization in ("modelopt_fp4", "modelopt_mixed") or quantization == "modelopt_fp4" ) - and is_cuda() - and (8, 0) <= get_device_capability() < (10, 0) + and get_platform().is_cuda + and (8, 0) <= get_platform().device_capability < (10, 0) ): overrides["moe_runner_backend"] = "marlin" logger.info( @@ -1568,10 +1551,10 @@ def _nemotron_h_overrides(server_args: Any, hf_config: Any) -> dict: else: overrides["moe_runner_backend"] = "flashinfer_cutlass" - if is_blackwell_supported() and is_attention_backend_not_set(cfg): + if get_platform().is_blackwell and is_attention_backend_not_set(cfg): if cfg.speculative_algorithm is not None: speculative_algorithm = cfg.speculative_algorithm.upper() - if is_sm100_supported() and cfg.speculative_eagle_topk in ( + if get_platform().is_sm100 and cfg.speculative_eagle_topk in ( None, 1, ): @@ -1592,7 +1575,7 @@ def _nemotron_h_overrides(server_args: Any, hf_config: Any) -> dict: and speculative_algorithm in ("EAGLE", "NEXTN", "DFLASH", "DSPARK") ): overrides["speculative_draft_attention_backend"] = "flashinfer" - elif is_sm100_supported(): + elif get_platform().is_sm100: overrides["attention_backend"] = "trtllm_mha" return overrides @@ -1606,7 +1589,7 @@ def _nemotron_h_overrides(server_args: Any, hf_config: Any) -> dict: ) def _qwen3_5_hybrid_overrides(server_args: Any, hf_config: Any) -> dict: cfg = resolving_view(server_args) - if not is_sm100_supported() or cfg.attention_backend is not None: + if not get_platform().is_sm100 or cfg.attention_backend is not None: return {} sm100_default_attn_backend = "triton" # trtllm_mha requires speculative_eagle_topk == 1 and page_size > 1. @@ -1647,7 +1630,11 @@ def _interns2_mobius_baseline_overrides(server_args: Any, hf_config: Any) -> dic def _qwen3vl_overrides(server_args: Any, hf_config: Any) -> dict: cfg = resolving_view(server_args) - if is_hip() and envs.SGLANG_USE_AITER_UNIFIED_ATTN.get() and cfg.page_size is None: + if ( + get_platform().is_hip + and envs.SGLANG_USE_AITER_UNIFIED_ATTN.get() + and cfg.page_size is None + ): logger.info( "Setting page_size=16 for aiter unified attention on Qwen3VLForConditionalGeneration." ) @@ -1666,7 +1653,7 @@ def _qwen3vl_overrides(server_args: Any, hf_config: Any) -> dict: def _qwen3_moe_family_overrides(server_args: Any, hf_config: Any) -> dict: cfg = resolving_view(server_args) overrides: Dict[str, Any] = {} - if is_sm100_supported(): + if get_platform().is_sm100: quant_method = get_quantization_config(hf_config) quantization = cfg.quantization if ( @@ -1693,7 +1680,7 @@ def _qwen3_moe_family_overrides(server_args: Any, hf_config: Any) -> dict: def _glm4_moe_overrides(server_args: Any, hf_config: Any) -> dict: cfg = resolving_view(server_args) overrides: Dict[str, Any] = {} - if is_sm100_supported(): + if get_platform().is_sm100: quantization_config = getattr(hf_config, "quantization_config", None) quant_method = ( quantization_config.get("quant_method") @@ -1734,9 +1721,9 @@ def _olmo2_overrides(server_args: Any, hf_config: Any) -> dict: ) overrides["disable_hybrid_swa_memory"] = True if cfg.attention_backend is None: - if is_cuda() and is_sm100_supported(): + if get_platform().is_cuda and get_platform().is_sm100: overrides["attention_backend"] = "trtllm_mha" - elif is_cuda() and get_device_sm() >= 80: + elif get_platform().is_cuda and get_platform().device_sm >= 80: overrides["attention_backend"] = "fa3" else: overrides["attention_backend"] = "triton" @@ -1751,10 +1738,10 @@ def _step3p_overrides(server_args: Any, hf_config: Any) -> dict: cfg = resolving_view(server_args) overrides: Dict[str, Any] = {} if is_attention_backend_not_set(cfg): - if is_blackwell_supported(): + if get_platform().is_blackwell: logger.info("Auto-select fa4 attention backend for Step3p7 on Blackwell.") overrides["attention_backend"] = "fa4" - elif is_sm90_supported(): + elif get_platform().is_sm90: logger.info("Auto-select fa3 attention backend for Step3p7 on Hopper.") overrides["attention_backend"] = "fa3" if cfg.speculative_algorithm == "EAGLE": @@ -1904,7 +1891,7 @@ def _dsa_kv_cache_dtype_default(view: Any) -> dict: return {} if not is_deepseek_dsa(hf_config): return {} - if is_npu() or is_xpu(): + if get_platform().is_npu or get_platform().is_xpu: return {} import torch @@ -1974,7 +1961,7 @@ def _dsa_split_backend_resolution(view: Any) -> dict: return {} if not is_deepseek_dsa(hf_config): return {} - if is_npu() or is_xpu(): + if get_platform().is_npu or get_platform().is_xpu: return {} import torch @@ -1989,7 +1976,7 @@ def _dsa_split_backend_resolution(view: Any) -> dict: model_arch == "GlmMoeDsaForCausalLM" and major == 12 and kv_cache_dtype == "fp8_e4m3" - and not is_hip() + and not get_platform().is_hip ) if is_glm_sm12_fp8: @@ -2020,7 +2007,7 @@ def _dsa_split_backend_resolution(view: Any) -> dict: ) return declared - if not user_set_prefill and not user_set_decode and is_hip(): + if not user_set_prefill and not user_set_decode and get_platform().is_hip: declared["dsa_prefill_backend"] = "tilelang" declared["dsa_decode_backend"] = "tilelang" elif kv_cache_dtype == "fp8_e4m3": @@ -2039,7 +2026,9 @@ def _dsa_split_backend_resolution(view: Any) -> dict: prefill = declared.get("dsa_prefill_backend", view.dsa_prefill_backend) decode = declared.get("dsa_decode_backend", view.dsa_decode_backend) - _check_tilelang_dsa_fp8_kv(kv_cache_dtype, prefill, decode, hip=is_hip()) + _check_tilelang_dsa_fp8_kv( + kv_cache_dtype, prefill, decode, hip=get_platform().is_hip + ) logger.warning( f"Set DSA backends for {kv_cache_dtype} KV Cache: " f"prefill={prefill}, decode={decode}." @@ -2074,7 +2063,7 @@ def _deepseek_moe_quant_resolution(view: Any) -> dict: if model_arch not in _DEEPSEEK_FAMILY_ARCHS: return {} overrides: Dict[str, Any] = {} - if is_sm100_supported(): + if get_platform().is_sm100: quant_method = get_quantization_config(hf_config) quant_cfg = getattr(hf_config, "quantization_config", None) or {} config_groups = quant_cfg.get("config_groups", {}) @@ -2160,7 +2149,7 @@ def _deepseek_spec_moe_resolution(view: Any) -> dict: model_arch = hf_config.architectures[0] if model_arch not in _DEEPSEEK_FAMILY_ARCHS: return {} - if not is_hip(): + if not get_platform().is_hip: return {} if not ( view.quantization == "modelopt_fp4" @@ -2225,7 +2214,7 @@ def _deepseek_v4_kv_cache_dtype(view: Any) -> dict: @_register_for("MuseGlimmerForConditionalGeneration", "MuseGlimmerForCausalLM") def _muse_glimmer_fp4_gemm_runner_overrides(server_args: Any, hf_config: Any) -> dict: cfg = resolving_view(server_args) - if is_sm120_supported() and cfg.fp4_gemm_runner_backend == "auto": + if get_platform().is_sm120 and cfg.fp4_gemm_runner_backend == "auto": logger.info("Use marlin as FP4 GEMM runner backend on SM120 for Muse Glimmer") return {"fp4_gemm_runner_backend": "marlin"} return {} @@ -2293,10 +2282,10 @@ def _flashinfer_allreduce_fusion_auto_enable(view: Any) -> dict: if ( view.flashinfer_allreduce_fusion_backend is None and model_arch in _FLASHINFER_ALLREDUCE_FUSION_ARCHS - and (is_sm90_supported() or is_sm100_supported()) + and (get_platform().is_sm90 or get_platform().is_sm100) and view.tp_size > 1 and not view.enable_dp_attention - and (view.nnodes == 1 or is_sm100_supported()) + and (view.nnodes == 1 or get_platform().is_sm100) and view.moe_a2a_backend == "none" ): logger.info( @@ -2324,7 +2313,7 @@ def _sampling_backend_default(view: Any) -> dict: if view.sampling_backend is None: return { "sampling_backend": ( - "flashinfer" if is_flashinfer_available() else "pytorch" + "flashinfer" if get_platform().has_flashinfer else "pytorch" ) } return {} @@ -2384,7 +2373,7 @@ def _deterministic_attention_backend(view: Any) -> dict: if view.attention_backend is None: # User didn't specify attention backend, fallback based on GPU architecture - if is_sm100_supported() or is_sm120_supported(): + if get_platform().is_sm100 or get_platform().is_sm120: # Blackwell and newer architectures if _deterministic_is_deepseek_model(view): # fallback to triton for DeepSeek models because flashinfer @@ -2517,7 +2506,7 @@ def _mla_kv_cache_dtype_checks(view: Any) -> dict: view.attention_backend == "trtllm_mla" or view.decode_attention_backend == "trtllm_mla" ): - if not is_blackwell_supported(): + if not get_platform().is_blackwell: raise ValueError( "TRTLLM MLA backend is only supported on Blackwell GPUs (SM100/SM12x). Please use a different backend." ) @@ -2529,7 +2518,7 @@ def _mla_kv_cache_dtype_checks(view: Any) -> dict: view.attention_backend == "tokenspeed_mla" or view.decode_attention_backend == "tokenspeed_mla" ): - if not is_blackwell_supported(): + if not get_platform().is_blackwell: raise ValueError( "tokenspeed_mla backend is only supported on Blackwell GPUs (SM100/SM12x)." ) @@ -2567,7 +2556,7 @@ def _cutedsl_prefill_backend_fill(view: Any) -> dict: assert ( view.prefill_attention_backend != "cutedsl_mla" ), "CuteDSL MLA only supports decoding for now" - if not is_sm100_supported(): + if not get_platform().is_sm100: raise ValueError( "CuteDSL MLA backend is only supported on Blackwell GPUs (SM100). Please use a different backend." ) @@ -2605,7 +2594,7 @@ def _fa4_page_constraint(view: Any) -> dict: or view.prefill_attention_backend == "fa4" ) and not use_mla_backend(view) - and is_sm100_supported() + and get_platform().is_sm100 # EAGLE topk>1 spec runs the two-pass page-tree cascade, which the FA4 # CUTLASS kernel aborts on at page_size>1. That path only works at # page_size==1, so skip the 128 auto-force for it and keep the default. @@ -2623,7 +2612,7 @@ def _attention_backend_platform_fallbacks(view: Any) -> dict: if ( view.attention_backend == "intel_amx" and view.device == "cpu" - and not cpu_has_amx_support() + and not get_platform().has_amx ): logger.warning( "The current platform does not support Intel AMX, will fallback to torch_native backend." @@ -2688,13 +2677,16 @@ def _page_size_default(view: Any) -> dict: # ROCm AITER backend, so the auto-bump is gated on HIP; on other # platforms the SHUFFLE 5D pool has no consumer kernels and the # env var is silently ignored (see MHATokenToKVPool). - if is_hip() and envs.SGLANG_AITER_KV_CACHE_LAYOUT.get().lower() == "vectorized_5d": + if ( + get_platform().is_hip + and envs.SGLANG_AITER_KV_CACHE_LAYOUT.get().lower() == "vectorized_5d" + ): logger.info( "Setting page_size=64 as default for " "SGLANG_AITER_KV_CACHE_LAYOUT=vectorized_5d." ) return {"page_size": 64} - if not is_musa(): + if not get_platform().is_musa: return {"page_size": 1} return {"page_size": 64} @@ -2765,7 +2757,7 @@ def _moe_runner_backend_quant_constraints(view: Any) -> dict: field) stay in the handler.""" moe_runner_backend = view.moe_runner_backend if view.quantization == "nvfp4_online": - if not is_sm100_supported(): + if not get_platform().is_sm100: raise ValueError( "--quantization nvfp4_online is supported only on " "NVIDIA Blackwell SM100/SM103 GPUs." @@ -2788,10 +2780,10 @@ def _moe_runner_backend_quant_constraints(view: Any) -> dict: # 128-alignment round-up off flashinfer_trtllm, so the experts would silently # load with gate and up exchanged. Leave the backend at "auto" and let # create_moe_runner resolve it to ASCEND. - if view.quantization == "mxfp8" and not is_npu(): + if view.quantization == "mxfp8" and not get_platform().is_npu: from sglang.srt.server_args import MXFP8_MOE_RUNNER_BACKEND_CHOICES - is_gfx95_mxfp8 = is_hip() and is_gfx95_supported() + is_gfx95_mxfp8 = get_platform().is_hip and is_gfx95_supported() allowed = list(MXFP8_MOE_RUNNER_BACKEND_CHOICES) if is_gfx95_mxfp8: allowed.append("triton") @@ -2808,7 +2800,7 @@ def _moe_runner_backend_quant_constraints(view: Any) -> dict: if ( moe_runner_backend == "auto" and view.quantization == "modelopt_fp4" - and is_sm120_supported() + and get_platform().is_sm120 ): moe_runner_backend = "flashinfer_cutlass" logger.info( @@ -2946,13 +2938,13 @@ def _gguf_quantization(view: Any) -> dict: def _dllm_attention_backend(view: Any) -> dict: if view.dllm_algorithm is None: return {} - if is_hip(): + if get_platform().is_hip: if view.attention_backend not in ["triton", "aiter"]: logger.warning( "Attention backend is set to triton for diffusion LLM inference on AMD GPUs" ) return {"attention_backend": "triton"} - elif is_npu(): + elif get_platform().is_npu: if view.attention_backend != "ascend": logger.warning( "Attention backend is overridden to 'ascend' when running on NPU for diffusion LLM inference." @@ -3094,7 +3086,7 @@ def get_default_attn_backend(server_args: Any, use_mla_backend: bool, model_conf if not use_mla_backend: # MHA architecture - if is_hopper_with_cuda_12_3() and is_no_spec_infer_or_topk_one( + if get_platform().is_hopper_with_cuda_12_3 and is_no_spec_infer_or_topk_one( resolved_view(server_args) ): # Note: flashinfer 0.6.1 caused performance regression on Hopper attention kernel @@ -3102,7 +3094,7 @@ def get_default_attn_backend(server_args: Any, use_mla_backend: bool, model_conf # ref: https://github.com/sgl-project/sglang/issues/17411 return "fa3" elif ( - is_sm100_supported() + get_platform().is_sm100 and is_no_spec_infer_or_topk_one(resolved_view(server_args)) and ( cfg.speculative_algorithm is None @@ -3114,22 +3106,22 @@ def get_default_attn_backend(server_args: Any, use_mla_backend: bool, model_conf if model_config.has_asymmetric_kv: return "fa4" return "trtllm_mha" - elif is_hip(): + elif get_platform().is_hip: return "aiter" elif is_mps(): return "torch_native" else: # FlashInfer does not support attention sinks. - if is_flashinfer_available() and not model_config.has_attention_sinks: + if get_platform().has_flashinfer and not model_config.has_attention_sinks: return "flashinfer" return "triton" else: # MLA architecture - if is_hopper_with_cuda_12_3(): + if get_platform().is_hopper_with_cuda_12_3: return "fa3" - elif is_sm100_supported(): + elif get_platform().is_sm100: return "flashinfer" - elif is_hip(): + elif get_platform().is_hip: head_num = model_config.get_num_kv_heads(cfg.tp_size) # TODO current aiter only support head number 16 or 128 head number if head_num == 128 or head_num == 16: diff --git a/python/sglang/srt/arg_groups/parallel_hook.py b/python/sglang/srt/arg_groups/parallel_hook.py index 31b19ee99..b066a56cb 100644 --- a/python/sglang/srt/arg_groups/parallel_hook.py +++ b/python/sglang/srt/arg_groups/parallel_hook.py @@ -21,7 +21,8 @@ from sglang.srt.arg_groups.overrides import ( from sglang.srt.connector import ConnectorType from sglang.srt.environ import envs from sglang.srt.model_executor.cuda_graph_config import Backend, Phase, with_phase -from sglang.srt.utils.common import is_cuda, parse_connector_type +from sglang.srt.runtime_context import get_platform +from sglang.srt.utils.common import parse_connector_type logger = logging.getLogger(__name__) @@ -139,7 +140,7 @@ def handle_dcp_validation(server_args: Any): "requires --dcp-size / --decode-context-parallel-size > 1, but " f"got dcp_size={cfg.dcp_size}." ) - if cfg.dcp_comm_backend == "fi_a2a" and not is_cuda(): + if cfg.dcp_comm_backend == "fi_a2a" and not get_platform().is_cuda: raise ValueError( "--dcp-comm-backend fi_a2a delegates the exchange to FlashInfer's " "MNNVL All-to-All kernel, which requires an NVIDIA CUDA platform " diff --git a/python/sglang/srt/arg_groups/platform_hook.py b/python/sglang/srt/arg_groups/platform_hook.py index 3ad779494..32d7028a0 100644 --- a/python/sglang/srt/arg_groups/platform_hook.py +++ b/python/sglang/srt/arg_groups/platform_hook.py @@ -12,7 +12,8 @@ from sglang.srt.arg_groups.overrides import ( ) from sglang.srt.hardware_backend.mlx.runtime import use_mlx from sglang.srt.model_executor.cuda_graph_config import Backend, Phase, with_phase -from sglang.srt.utils.common import is_cuda, is_hip, is_host_cpu_arm64, is_npu +from sglang.srt.runtime_context import get_platform +from sglang.srt.utils.common import is_host_cpu_arm64 logger = logging.getLogger(__name__) @@ -59,7 +60,7 @@ def handle_mps_backends(server_args: Any): def handle_amd_specifics(server_args: Any): - if is_hip(): + if get_platform().is_hip: declare_resolution( server_args, "_handle_amd_specifics", triton_attention_num_kv_splits=16 ) @@ -68,7 +69,9 @@ def handle_amd_specifics(server_args: Any): def handle_nccl_pre_warm(server_args: Any): # pre_warm_nccl is only used with CUDA or HIP hardware or NPU hardware cfg = resolving_view(server_args) - if cfg.pre_warm_nccl and not (is_cuda() or is_hip() or is_npu()): + if cfg.pre_warm_nccl and not ( + get_platform().is_cuda or get_platform().is_hip or get_platform().is_npu + ): logger.warning( "pre_warm_nccl is only applicable for CUDA or HIP hardware or NPU hardware. " "Ignoring pre_warm_nccl setting on current hardware." @@ -80,7 +83,7 @@ def handle_symm_mem_device_support(server_args: Any): cfg = resolving_view(server_args) # The symm-mem allocator compiles a CUDA plugin and links -lnccl, so off # CUDA/HIP (e.g. Ascend NPU) it fails deep in a build step rather than here. - if cfg.enable_symm_mem and not (is_cuda() or is_hip()): + if cfg.enable_symm_mem and not (get_platform().is_cuda or get_platform().is_hip): logger.warning( "--enable-symm-mem is not supported on non CUDA/HIP devices " "(NCCL symmetric memory is unavailable). Disabling symmetric memory." diff --git a/python/sglang/srt/arg_groups/serving_hook.py b/python/sglang/srt/arg_groups/serving_hook.py index 8219156ca..2cf7302f2 100644 --- a/python/sglang/srt/arg_groups/serving_hook.py +++ b/python/sglang/srt/arg_groups/serving_hook.py @@ -18,16 +18,11 @@ from sglang.srt.arg_groups.overrides import ( ) from sglang.srt.environ import envs from sglang.srt.model_executor.cuda_graph_config import Backend, Phase, with_phase +from sglang.srt.runtime_context import get_platform from sglang.srt.utils.common import ( configure_media_url_security, get_device, - get_device_sm, - is_cuda, - is_hip, is_mnnvl_fabric_device, - is_sm90_supported, - is_sm100_supported, - is_sm120_supported, ) from sglang.utils import is_in_ci @@ -409,7 +404,7 @@ def handle_environment_variables(server_args: Any): if cfg.enable_deterministic_inference: envs.SGLANG_FLASHINFER_MOE_FUSED_FINALIZE.set("0") if cfg.debug_cuda_graph: - if not (is_cuda() or is_hip()): + if not (get_platform().is_cuda or get_platform().is_hip): logger.warning( "--debug-cuda-graph is not supported on non CUDA/HIP devices. " "Disabling breakable CUDA graph." @@ -424,7 +419,7 @@ def handle_environment_variables(server_args: Any): "All operations will run eagerly through the graph capture/replay path." ) if cfg.enable_deepseek_v4_fp4_indexer and not ( - is_sm100_supported() or is_sm120_supported() + get_platform().is_sm100 or get_platform().is_sm120 ): raise ValueError( "--enable-deepseek-v4-fp4-indexer requires SM100 or SM120 GPUs with " @@ -434,13 +429,15 @@ def handle_environment_variables(server_args: Any): # it, mirroring the forward scale split: the ue8m0 path # (DEEPGEMM_SCALE_UE8M0, true sm100, default on) or an sm90 opt-in # fp32-scale path (use FP4 expert ckpt). Disable in every other case. - if is_cuda() and envs.SGLANG_OPT_FP8_WO_A_GEMM.get(): + if get_platform().is_cuda and envs.SGLANG_OPT_FP8_WO_A_GEMM.get(): from sglang.srt.layers import deep_gemm_wrapper - sm = get_device_sm() + sm = get_platform().device_sm explicit = envs.SGLANG_OPT_FP8_WO_A_GEMM.is_set() supported = deep_gemm_wrapper.DEEPGEMM_SCALE_UE8M0 or ( - deep_gemm_wrapper.ENABLE_JIT_DEEPGEMM and is_sm90_supported() and explicit + deep_gemm_wrapper.ENABLE_JIT_DEEPGEMM + and get_platform().is_sm90 + and explicit ) if not supported and explicit: logger.warning( @@ -788,7 +785,7 @@ def handle_multimodal_feature_transport(server_args: Any): ) elif ( model_config_of(server_args).is_multimodal - and is_cuda() + and get_platform().is_cuda and cfg.disaggregation_mode == "null" ): # A full GPU pool always degrades to CPU transport per tensor. @@ -849,7 +846,7 @@ def handle_multimodal_feature_transport(server_args: Any): requested_transport = "cpu" if requested_transport == "cuda_vmm": - if not is_cuda(): + if not get_platform().is_cuda: raise ValueError("--mm-feature-transport=cuda_vmm requires NVIDIA CUDA.") if cfg.pp_size != 1: raise ValueError( @@ -875,7 +872,7 @@ def handle_multimodal_feature_transport(server_args: Any): ) if requested_transport == "cuda_ipc": - if not is_cuda(): + if not get_platform().is_cuda: raise ValueError("--mm-feature-transport=cuda_ipc requires NVIDIA CUDA.") if cfg.nnodes != 1: raise ValueError( diff --git a/python/sglang/srt/arg_groups/speculative_hook.py b/python/sglang/srt/arg_groups/speculative_hook.py index 8a2650b63..7083af6c1 100644 --- a/python/sglang/srt/arg_groups/speculative_hook.py +++ b/python/sglang/srt/arg_groups/speculative_hook.py @@ -7,6 +7,7 @@ from typing import TYPE_CHECKING, Optional from sglang.srt.arg_groups.overrides import ( _speculative_moe_runner_default, + attention_backends_of, declare_direct_writes, declare_resolution, model_config_of, @@ -14,6 +15,7 @@ from sglang.srt.arg_groups.overrides import ( resolving_view, run_post_process_pass, ) +from sglang.srt.runtime_context import get_platform if TYPE_CHECKING: from sglang.srt.server_args import ServerArgs @@ -563,7 +565,6 @@ def _resolve_dflash_draft_attention_backend(server_args: ServerArgs) -> None: draft modes). """ cfg = resolving_view(server_args) - from sglang.srt.utils import is_hip supported_draft_backends = ( "flashinfer", @@ -574,11 +575,10 @@ def _resolve_dflash_draft_attention_backend(server_args: ServerArgs) -> None: "ascend", ) # Use triton on ROCm (no FlashInfer), flashinfer on CUDA. - fallback_backend = "triton" if is_hip() else "flashinfer" + fallback_backend = "triton" if get_platform().is_hip else "flashinfer" draft_backend = cfg.speculative_draft_attention_backend if draft_backend is None: - from sglang.srt.arg_groups.overrides import attention_backends_of draft_backend, _ = attention_backends_of(resolved_view(server_args)) if draft_backend is None: @@ -660,7 +660,6 @@ def _handle_frozen_kv_mtp(server_args: ServerArgs) -> None: def _handle_eagle_family(server_args: ServerArgs) -> None: cfg = resolving_view(server_args) - from sglang.srt.arg_groups.overrides import attention_backends_of if ( cfg.speculative_algorithm == "STANDALONE" diff --git a/python/sglang/srt/arg_groups/validation_hook.py b/python/sglang/srt/arg_groups/validation_hook.py index 6364f7ec9..e70af6594 100644 --- a/python/sglang/srt/arg_groups/validation_hook.py +++ b/python/sglang/srt/arg_groups/validation_hook.py @@ -17,7 +17,8 @@ from sglang.srt.arg_groups.overrides import ( from sglang.srt.distributed.device_communicators.mooncake_transfer_engine import ( parse_ib_device_config, ) -from sglang.srt.utils.common import is_hip, is_npu, torch_release +from sglang.srt.runtime_context import get_platform +from sglang.srt.utils.common import torch_release from sglang.srt.utils.runai_utils import is_runai_obj_uri logger = logging.getLogger(__name__) @@ -164,7 +165,9 @@ def check_server_args(server_args: Any): ), "schedule_conservativeness must be non-negative" if cfg.model_impl == "mindspore": - assert is_npu(), "MindSpore model impl is only supported on Ascend npu." + assert ( + get_platform().is_npu + ), "MindSpore model impl is only supported on Ascend npu." # Check metrics labels if ( @@ -410,7 +413,7 @@ def check_two_batch_overlap(server_args: Any): cfg = resolving_view(server_args) cp_tbo = ( - is_hip() + get_platform().is_hip and cfg.enable_dsa_prefill_context_parallel and cfg.dsa_prefill_cp_mode == "round-robin-split" ) diff --git a/python/sglang/srt/configs/inkling.py b/python/sglang/srt/configs/inkling.py index c963ca9e9..7c1859f1f 100644 --- a/python/sglang/srt/configs/inkling.py +++ b/python/sglang/srt/configs/inkling.py @@ -8,7 +8,10 @@ from transformers import CONFIG_MAPPING from transformers.configuration_utils import PretrainedConfig from sglang.srt.configs.mamba_utils import BaseLinearStateParams -from sglang.srt.runtime_context import get_exec +from sglang.srt.runtime_context import ( + get_exec, + get_parallel, +) class InklingModelConfig(PretrainedConfig): @@ -210,7 +213,6 @@ class InklingModelConfig(PretrainedConfig): @property def mamba2_cache_params(self) -> Optional[InklingConvCacheParams]: - from sglang.srt.runtime_context import get_parallel try: tp_size = get_parallel().attn_tp_size diff --git a/python/sglang/srt/configs/model_config.py b/python/sglang/srt/configs/model_config.py index 32ea0fd3c..1785a7c8d 100644 --- a/python/sglang/srt/configs/model_config.py +++ b/python/sglang/srt/configs/model_config.py @@ -30,8 +30,9 @@ from sglang.srt.configs.embedding_model_spec import resolve_embedding_model_spec from sglang.srt.configs.linear_attn_model_registry import get_linear_attn_config from sglang.srt.environ import envs from sglang.srt.layers.quantization import QUANTIZATION_METHODS +from sglang.srt.runtime_context import get_platform from sglang.srt.server_args import ServerArgs -from sglang.srt.utils import is_hip, is_sm100_supported, retry +from sglang.srt.utils import is_hip, retry from sglang.srt.utils.hf_transformers_utils import ( get_config, get_context_length, @@ -1656,7 +1657,7 @@ class ModelConfig: if self.quantization not in optimized_quantization_methods: # Don't warn for MXFP4/MXFP8 on SM100 since they have optimized kernels if not ( - self.quantization in ["mxfp4", "mxfp8"] and is_sm100_supported() + self.quantization in ["mxfp4", "mxfp8"] and get_platform().is_sm100 ): logger.warning( "%s quantization is not fully " diff --git a/python/sglang/srt/debug_utils/dumper.py b/python/sglang/srt/debug_utils/dumper.py index 6bdef9ebd..906b980c1 100644 --- a/python/sglang/srt/debug_utils/dumper.py +++ b/python/sglang/srt/debug_utils/dumper.py @@ -23,7 +23,11 @@ import torch.distributed as dist import zmq from sglang.srt.managers.io_struct import sock_recv, sock_send, wrap_as_pickle -from sglang.srt.runtime_context import get_serving +from sglang.srt.runtime_context import ( + get_parallel, + get_server_args, + get_serving, +) # -------------------------------------- config base ------------------------------------------ @@ -1722,8 +1726,6 @@ class _SGLangPlugin(_FrameworkPlugin): info = {} - from sglang.srt.runtime_context import get_parallel - try: parallel = get_parallel() info["tp_rank"] = parallel.tp_rank @@ -1793,7 +1795,6 @@ class _SGLangPlugin(_FrameworkPlugin): return None try: - from sglang.srt.runtime_context import get_server_args args = get_server_args() if args is None: diff --git a/python/sglang/srt/disaggregation/encoder/preprocessor.py b/python/sglang/srt/disaggregation/encoder/preprocessor.py index bd6d3f06e..f05229d8b 100644 --- a/python/sglang/srt/disaggregation/encoder/preprocessor.py +++ b/python/sglang/srt/disaggregation/encoder/preprocessor.py @@ -168,7 +168,7 @@ class EncoderPreprocessor: self.image_processor = AutoImageProcessor.from_pretrained( get_serving().tokenizer_path or get_model().model_path, trust_remote_code=get_model().trust_remote_code, - revision=server_args.revision, + revision=get_model().revision, **image_processor_kwargs, ) except Exception as e: @@ -179,7 +179,7 @@ class EncoderPreprocessor: self.video_processor = AutoVideoProcessor.from_pretrained( get_serving().tokenizer_path or get_model().model_path, trust_remote_code=get_model().trust_remote_code, - revision=server_args.revision, + revision=get_model().revision, ) except Exception as e: logger.warning(f"Failed to load video processor: {e}") @@ -189,7 +189,7 @@ class EncoderPreprocessor: _audio_proc = AutoProcessor.from_pretrained( get_serving().tokenizer_path or get_model().model_path, trust_remote_code=get_model().trust_remote_code, - revision=server_args.revision, + revision=get_model().revision, ) if not hasattr(_audio_proc, "feature_extractor"): logger.warning( diff --git a/python/sglang/srt/disaggregation/encoder/receiver.py b/python/sglang/srt/disaggregation/encoder/receiver.py index 53bcff1fc..fd5a8564d 100644 --- a/python/sglang/srt/disaggregation/encoder/receiver.py +++ b/python/sglang/srt/disaggregation/encoder/receiver.py @@ -1736,7 +1736,7 @@ class MMReceiverBase(ABC): self.tp_rank = tp_rank self.tp_size = get_parallel().tp_size self.tp_group = tp_group - self.nnodes = server_args.nnodes + self.nnodes = get_parallel().nnodes self.hostname = get_local_ip_auto() self.waiting_list: List[WaitingMMRequestBase] = [] self.waiting_by_rid: Dict[str, WaitingMMRequestBase] = {} @@ -1838,19 +1838,19 @@ class MMReceiverBase(ABC): extra_kwargs = {} if getattr(server_args, "tokenizer_backend", None) is not None: - extra_kwargs["tokenizer_backend"] = server_args.tokenizer_backend + extra_kwargs["tokenizer_backend"] = get_serving().tokenizer_backend _processor = get_processor( get_serving().tokenizer_path, - tokenizer_mode=server_args.tokenizer_mode, + tokenizer_mode=get_serving().tokenizer_mode, trust_remote_code=get_model().trust_remote_code, - revision=server_args.revision, + revision=get_model().revision, image_processor_backend=resolve_image_processor_backend(get_mm()), **extra_kwargs, ) enable_adaptive_dispatch_to_encoder = ( - server_args.enable_adaptive_dispatch_to_encoder + get_disagg().enable_adaptive_dispatch_to_encoder ) mm_processor_kwargs = {} if model_config is not None: diff --git a/python/sglang/srt/disaggregation/mooncake/conn.py b/python/sglang/srt/disaggregation/mooncake/conn.py index 202d014e2..9d93721db 100644 --- a/python/sglang/srt/disaggregation/mooncake/conn.py +++ b/python/sglang/srt/disaggregation/mooncake/conn.py @@ -63,7 +63,11 @@ from sglang.srt.observability.trace import ( TraceReqContext, trace_set_thread_info, ) -from sglang.srt.runtime_context import get_memory, get_schedule +from sglang.srt.runtime_context import ( + get_memory, + get_observability, + get_schedule, +) from sglang.srt.server_args import ServerArgs from sglang.srt.utils.network import NetworkAddress @@ -213,7 +217,7 @@ class MooncakeKVManager(StagingManagerMixin, CommonKVManager): self.init_engine() self.register_buffer_to_engine() self.enable_staging = envs.SGLANG_DISAGG_STAGING_BUFFER.get() - self.enable_trace = server_args.enable_trace + self.enable_trace = get_observability().enable_trace if self.disaggregation_mode == DisaggregationMode.PREFILL: self.start_prefill_thread() self.session_failures = defaultdict(int) diff --git a/python/sglang/srt/distributed/bootstrap.py b/python/sglang/srt/distributed/bootstrap.py index d1d0f52bf..570d6ba98 100644 --- a/python/sglang/srt/distributed/bootstrap.py +++ b/python/sglang/srt/distributed/bootstrap.py @@ -29,6 +29,7 @@ from sglang.srt.environ import envs from sglang.srt.layers.dp_attention import initialize_dp_attention from sglang.srt.platforms import current_platform from sglang.srt.runtime_context import ( + get_disagg, get_exec, get_parallel, get_serving, @@ -82,12 +83,10 @@ def init_torch_distributed( backend = _resolve_backend(device=device, server_args=server_args) before_avail_memory = get_available_gpu_memory(device, ps.gpu_id) - if not server_args.enable_p2p_check: + if not get_parallel().enable_p2p_check: monkey_patch_p2p_access_check() - dist_init_method = _resolve_dist_init_method( - server_args=server_args, dist_port=dist_port - ) + dist_init_method = _resolve_dist_init_method(dist_port=dist_port) _set_all_reduce_flags(server_args=server_args) if not is_draft_worker: @@ -135,7 +134,7 @@ def init_torch_distributed( _prewarm_tp_lm_head_all_to_all() maybe_wait_for_gated_launch( - host=server_args.host, port=server_args.gated_launch_port + host=get_serving().host, port=get_parallel().gated_launch_port ) # Draft workers reuse the target pool config and may exist on only one PP stage; @@ -177,7 +176,7 @@ def _resolve_backend(*, device: str, server_args: ServerArgs) -> str: return backend -def _resolve_dist_init_method(*, server_args: ServerArgs, dist_port: int) -> str: +def _resolve_dist_init_method(*, dist_port: int) -> str: # Allow external orchestrators (e.g. trainpi) to override the distributed # init method. When set to "env://", torch uses MASTER_ADDR/MASTER_PORT # env-vars and an externally-created TCPStore, completely avoiding port @@ -185,8 +184,8 @@ def _resolve_dist_init_method(*, server_args: ServerArgs, dist_port: int) -> str dist_init_method_override = envs.SGLANG_DISTRIBUTED_INIT_METHOD_OVERRIDE.get() if dist_init_method_override: dist_init_method = dist_init_method_override - elif server_args.dist_init_addr: - na = NetworkAddress.parse(server_args.dist_init_addr) + elif get_parallel().dist_init_addr: + na = NetworkAddress.parse(get_parallel().dist_init_addr) dist_init_method = na.to_tcp() else: dist_init_method = NetworkAddress( @@ -240,7 +239,7 @@ def _init_parallel_groups( ) -> None: is_ep_joiner = server_args.is_ep_joiner is_scale_joiner = server_args.is_ep_scale_joiner - rank_offset = server_args.ep_join_rank_offset if is_scale_joiner else 0 + rank_offset = get_parallel().ep_join_rank_offset if is_scale_joiner else 0 world_size = ( rank_offset + tp_size * pp_size if is_scale_joiner else tp_size * pp_size ) @@ -252,10 +251,10 @@ def _init_parallel_groups( rank=rank, local_rank=gpu_id, distributed_init_method=dist_init_method, - timeout=server_args.dist_timeout, + timeout=get_parallel().dist_timeout, moe_a2a_backend=get_exec().moe.moe_a2a_backend, recovered_rank=is_ep_joiner, - max_world_size=server_args.max_ep_size, + max_world_size=get_parallel().max_ep_size, ) initialize_model_parallel( tensor_model_parallel_size=tp_size, @@ -265,7 +264,7 @@ def _init_parallel_groups( attention_context_model_parallel_size=attn_cp_size, moe_data_model_parallel_size=moe_dp_size, decode_context_parallel_size=dcp_size, - duplicate_tp_group=server_args.enable_pdmux, + duplicate_tp_group=get_disagg().enable_pdmux, duplicate_attn_cp_group=( is_hip() and server_args.enable_two_batch_overlap @@ -274,7 +273,7 @@ def _init_parallel_groups( enable_symm_mem=get_exec().comm.enable_symm_mem, recovered_rank=is_ep_joiner, rank_offset=rank_offset, - max_world_size=server_args.max_ep_size, + max_world_size=get_parallel().max_ep_size, ) _tag_groups_for_flashinfer_allreduce_only() initialize_dp_attention( diff --git a/python/sglang/srt/distributed/device_communicators/triton_symm_mem_ag.py b/python/sglang/srt/distributed/device_communicators/triton_symm_mem_ag.py index 0ec0b123c..fa508304c 100644 --- a/python/sglang/srt/distributed/device_communicators/triton_symm_mem_ag.py +++ b/python/sglang/srt/distributed/device_communicators/triton_symm_mem_ag.py @@ -16,6 +16,12 @@ import torch.distributed._symmetric_memory as symm_mem import triton import triton.language as tl +from sglang.srt.runtime_context import ( + get_parallel, + get_schedule, + get_spec, +) + logger = logging.getLogger(__name__) # Each thread moves _NUMEL_PER_THREAD bf16 via one 128-bit multimem op; the @@ -423,7 +429,6 @@ def recommended_max_tokens(include_prefill: bool, floor: int = 0) -> int: NCCL. Covers the spec-decode batch plus, if ``include_prefill``, a prefill chunk. Returns ``floor`` if server args are unavailable.""" try: - from sglang.srt.runtime_context import get_schedule, get_spec def g(value) -> int: return value if isinstance(value, int) and value > 0 else 0 @@ -466,7 +471,6 @@ class MultimemAllGatherer: # Lazy import avoids a module-load dependency on the distributed facade. from sglang.srt.distributed import get_tp_group from sglang.srt.distributed.parallel_state import in_the_same_node_as - from sglang.srt.runtime_context import get_parallel tp_group = get_tp_group() # Only probe node topology when the deployment can actually span diff --git a/python/sglang/srt/distributed/parallel_state.py b/python/sglang/srt/distributed/parallel_state.py index c66ee20ab..4d7c64e40 100644 --- a/python/sglang/srt/distributed/parallel_state.py +++ b/python/sglang/srt/distributed/parallel_state.py @@ -2869,7 +2869,6 @@ def patch_tensor_parallel_group(tp_group: GroupCoordinator): Args: tp_group (GroupCoordinator): the tp group coordinator """ - from sglang.srt.runtime_context import get_parallel global _TP_STATE_PATCHED assert not _TP_STATE_PATCHED, "Should not call when it's already patched" diff --git a/python/sglang/srt/distributed/utils.py b/python/sglang/srt/distributed/utils.py index 9829b086d..0aeead003 100644 --- a/python/sglang/srt/distributed/utils.py +++ b/python/sglang/srt/distributed/utils.py @@ -17,13 +17,14 @@ from typing import Any, Deque, Dict, Optional, Sequence, Tuple import torch from torch.distributed import TCPStore +from sglang.srt.runtime_context import get_resources + logger = logging.getLogger(__name__) def set_global_tcp_store(store: TCPStore) -> None: """Install the shared TCPStore created during distributed initialization; the handle lives on ``ctx.resources``.""" - from sglang.srt.runtime_context import get_resources get_resources().tcp_store = store logger.info("Global TCPStore has been set") @@ -39,7 +40,6 @@ def get_global_tcp_store() -> Optional[TCPStore]: Returns: The global TCPStore instance, or None if not initialized yet. """ - from sglang.srt.runtime_context import get_resources store = get_resources().tcp_store if store is None: diff --git a/python/sglang/srt/dllm/mixin/scheduler.py b/python/sglang/srt/dllm/mixin/scheduler.py index abc4b3ed5..86b1a9500 100644 --- a/python/sglang/srt/dllm/mixin/scheduler.py +++ b/python/sglang/srt/dllm/mixin/scheduler.py @@ -319,7 +319,7 @@ class SchedulerDllmMixin: if running_batch.batch_is_full: if ( not self.enable_priority_preemption - or not adder.preempt_to_schedule(req, self.server_args) + or not adder.preempt_to_schedule(req) ): break diff --git a/python/sglang/srt/entrypoints/engine.py b/python/sglang/srt/entrypoints/engine.py index a970a050c..55a22a251 100644 --- a/python/sglang/srt/entrypoints/engine.py +++ b/python/sglang/srt/entrypoints/engine.py @@ -49,6 +49,7 @@ import uvloop import zmq from sglang.srt.arg_groups.overrides import ( + attention_backends_of, resolved_view, resolving_view, ) @@ -1630,7 +1631,6 @@ class Engine(EngineScoreMixin, EngineBase): def _set_envs_and_config(server_args: ServerArgs): - from sglang.srt.arg_groups.overrides import attention_backends_of cfg = resolving_view(server_args) # Set global environments diff --git a/python/sglang/srt/entrypoints/grpc_bridge.py b/python/sglang/srt/entrypoints/grpc_bridge.py index 4adf1d737..853c1c62e 100644 --- a/python/sglang/srt/entrypoints/grpc_bridge.py +++ b/python/sglang/srt/entrypoints/grpc_bridge.py @@ -17,6 +17,7 @@ from pydantic import ValidationError from sglang.srt.arg_groups.overrides import resolving_view from sglang.srt.configs.embedding_model_spec import resolved_embedding_plan from sglang.srt.runtime_context import ( + describe_kv_events_publisher, get_lora, get_serving, ) @@ -425,8 +426,8 @@ class RuntimeHandle: def get_server_info(self) -> str: result: Dict[str, Any] = self.tokenizer_manager.server_args.resolved_dict() result.update(self.scheduler_info) - result["kv_events"] = ( - self.tokenizer_manager.server_args.describe_kv_events_publisher() + result["kv_events"] = describe_kv_events_publisher( + self.tokenizer_manager.server_args ) return json.dumps(msgspec_to_builtins(result), default=str) diff --git a/python/sglang/srt/entrypoints/http_server.py b/python/sglang/srt/entrypoints/http_server.py index 2d5c6ffbb..fa76cbe12 100644 --- a/python/sglang/srt/entrypoints/http_server.py +++ b/python/sglang/srt/entrypoints/http_server.py @@ -409,7 +409,7 @@ async def lifespan(fast_api_app: FastAPI): if server_args.sidecar is not None: from sglang.srt.entrypoints.sidecar import start_sidecar - sidecar = start_sidecar(server_args) + sidecar = start_sidecar() # Execute the general warmup warmup_thread = threading.Thread( @@ -485,6 +485,7 @@ app.include_router(v1_loads_router) from sglang.srt.entrypoints.elastic_ep import router as elastic_ep_router from sglang.srt.runtime_context import ( + describe_kv_events_publisher, get_disagg, get_exec, get_lora, @@ -832,8 +833,8 @@ async def server_info(): "version": __version__, # Structured KV-event publisher descriptor for KV-aware routers. # `None` when publishing is disabled or misconfigured; see - # `ServerArgs.describe_kv_events_publisher` for the precise contract. - "kv_events": server_args.describe_kv_events_publisher(), + # `runtime_context.describe_kv_events_publisher` for the contract. + "kv_events": describe_kv_events_publisher(server_args), } ) diff --git a/python/sglang/srt/entrypoints/sidecar.py b/python/sglang/srt/entrypoints/sidecar.py index 754950af2..0b1d59c5f 100644 --- a/python/sglang/srt/entrypoints/sidecar.py +++ b/python/sglang/srt/entrypoints/sidecar.py @@ -114,10 +114,10 @@ class Sidecar: kill_process_tree(self.proc.pid, wait_timeout=self.shutdown_timeout) -def start_sidecar(server_args) -> Sidecar: - module_name = server_args.sidecar +def start_sidecar() -> Sidecar: + module_name = get_serving().sidecar assert module_name is not None - sidecar_args, shutdown_timeout = _parse_sidecar_args(server_args.sidecar_args) + sidecar_args, shutdown_timeout = _parse_sidecar_args(get_serving().sidecar_args) endpoint = build_sidecar_endpoint(get_serving().host, get_serving().grpc_port) proc = mp.get_context("spawn").Process( name=f"sglang_sidecar_{module_name}", diff --git a/python/sglang/srt/eplb/expert_location.py b/python/sglang/srt/eplb/expert_location.py index b2ced8192..c1d16e276 100644 --- a/python/sglang/srt/eplb/expert_location.py +++ b/python/sglang/srt/eplb/expert_location.py @@ -28,6 +28,8 @@ import torch.nn.functional as F from sglang.srt.runtime_context import ( get_device, get_exec, + get_parallel, + get_resources, ) if TYPE_CHECKING: @@ -38,7 +40,6 @@ logger = logging.getLogger(__name__) def _prefer_same_node_experts() -> bool: from sglang.srt.elastic_ep.elastic_ep import elastic_expanded_world_enabled - from sglang.srt.runtime_context import get_exec return ( get_exec().moe.ep_join_mode != "scale" and not elastic_expanded_world_enabled() @@ -185,8 +186,6 @@ class ExpertLocationMetadata: logical_count = logical_count.unsqueeze(0) logical_count = logical_count.to(get_device().device) - from sglang.srt.runtime_context import get_parallel - common = ExpertLocationMetadata._init_common(model_config) if common is None: @@ -224,7 +223,6 @@ class ExpertLocationMetadata: @staticmethod def _init_common(model_config: ModelConfig): - from sglang.srt.runtime_context import get_exec, get_parallel model_config_for_expert_location = ( ModelConfigForExpertLocation.from_model_config(model_config) @@ -273,7 +271,6 @@ class ExpertLocationMetadata: logical_to_all_physical_map: torch.Tensor, moe_ep_rank: Optional[int] = None, ): - from sglang.srt.runtime_context import get_exec _, num_physical_experts = physical_to_logical_map.shape @@ -481,13 +478,11 @@ def _normalize_layer_ids( def get_global_expert_location_metadata(): - from sglang.srt.runtime_context import get_resources return get_resources().expert_location_metadata def set_global_expert_location_metadata(value, allow_overwrite=False): - from sglang.srt.runtime_context import get_resources resources = get_resources() if not allow_overwrite: @@ -541,7 +536,6 @@ def _compute_logical_to_all_physical_map( ep_size: int, moe_ep_rank: int, ): - from sglang.srt.runtime_context import get_exec, get_parallel # This is rarely called, so we use for loops for maximum clarity @@ -623,7 +617,6 @@ def compute_logical_to_rank_dispatch_physical_map( ep_rank: int, seed: int = 42, ): - from sglang.srt.runtime_context import get_parallel r = random.Random(seed) diff --git a/python/sglang/srt/eplb/lplb_solver.py b/python/sglang/srt/eplb/lplb_solver.py index 412641dcc..31b24933d 100644 --- a/python/sglang/srt/eplb/lplb_solver.py +++ b/python/sglang/srt/eplb/lplb_solver.py @@ -23,6 +23,8 @@ from typing import Optional import torch +from sglang.srt.runtime_context import get_resources + logger = logging.getLogger(__name__) # Global per-layer LPLB solvers @@ -58,19 +60,16 @@ def assert_lplb_supported_model(architecture: str) -> None: def get_global_lplb_solver(layer_id: int) -> Optional[LPLBSolver]: - from sglang.srt.runtime_context import get_resources return get_resources().lplb_solvers.get(layer_id) def set_global_lplb_solver(layer_id: int, solver: LPLBSolver): - from sglang.srt.runtime_context import get_resources get_resources().lplb_solvers[layer_id] = solver def clear_global_lplb_solvers(): - from sglang.srt.runtime_context import get_resources get_resources().lplb_solvers.clear() diff --git a/python/sglang/srt/hardware_backend/npu/attention/ascend_backend.py b/python/sglang/srt/hardware_backend/npu/attention/ascend_backend.py index 88e6ce4f7..0d90b5bc9 100644 --- a/python/sglang/srt/hardware_backend/npu/attention/ascend_backend.py +++ b/python/sglang/srt/hardware_backend/npu/attention/ascend_backend.py @@ -26,7 +26,11 @@ from sglang.srt.layers.utils.cp_utils import cp_all_gather_rerange_kv_cache from sglang.srt.mem_cache.memory_pool import KVWriteLoc from sglang.srt.mem_cache.swa_memory_pool import SWAKVPool from sglang.srt.model_executor.forward_batch_info import ForwardBatch, ForwardMode -from sglang.srt.runtime_context import get_flags, get_spec +from sglang.srt.runtime_context import ( + get_flags, + get_parallel, + get_spec, +) from sglang.srt.speculative.spec_info import SpecInput, SpecInputType from sglang.srt.utils import ( get_bool_env_var, @@ -42,8 +46,6 @@ import logging import numpy as np -from sglang.srt.runtime_context import get_parallel - logger = logging.getLogger(__name__) FULL_ATTENTION_WINDOW = 2147483647 diff --git a/python/sglang/srt/layers/attention/attention_registry.py b/python/sglang/srt/layers/attention/attention_registry.py index a0615766d..d60c4de0b 100644 --- a/python/sglang/srt/layers/attention/attention_registry.py +++ b/python/sglang/srt/layers/attention/attention_registry.py @@ -2,7 +2,10 @@ import logging import warnings from typing import TYPE_CHECKING -from sglang.srt.arg_groups.overrides import resolved_view +from sglang.srt.arg_groups.overrides import ( + attention_backends_of, + resolved_view, +) from sglang.srt.configs.hybrid_arch import ( hybrid_gdn_config, hybrid_lightning_config, @@ -16,6 +19,7 @@ from sglang.srt.configs.linear_attn_model_registry import ( ) from sglang.srt.runtime_context import ( get_parallel, + get_platform, get_spec, ) from sglang.srt.utils import get_device_capability, is_hip, is_musa, is_npu @@ -75,7 +79,6 @@ def create_trtllm_mla_backend(runner): if not runner.use_mla_backend: raise ValueError("trtllm_mla backend can only be used with MLA models.") if get_parallel().dcp_enabled and get_spec().speculative_algorithm is not None: - from sglang.srt.arg_groups.overrides import attention_backends_of _, decode_backend = attention_backends_of(resolved_view(runner.server_args)) if decode_backend == "trtllm_mla": @@ -394,7 +397,6 @@ def attn_backend_wrapper(runner: "ModelRunner", full_attn_backend: "AttentionBac from sglang.srt.utils import ( is_blackwell, is_npu, - is_sm120_supported, is_xpu, ) @@ -433,7 +435,7 @@ def attn_backend_wrapper(runner: "ModelRunner", full_attn_backend: "AttentionBac hybrid_backend_cls = HybridLinearAttnBackend if hybrid_gdn_config(runner.model_config) is not None: if is_blackwell(): - if is_sm120_supported(): + if get_platform().is_sm120: allowed = {"triton", "trtllm_mha", "flashinfer"} else: allowed = {"triton", "trtllm_mha", "fa4"} diff --git a/python/sglang/srt/layers/attention/deepseek_v4_backend.py b/python/sglang/srt/layers/attention/deepseek_v4_backend.py index 315496a16..2f1f1f2f3 100644 --- a/python/sglang/srt/layers/attention/deepseek_v4_backend.py +++ b/python/sglang/srt/layers/attention/deepseek_v4_backend.py @@ -71,6 +71,7 @@ from sglang.srt.mem_cache.deepseek_v4_memory_pool import DeepSeekV4TokenToKVPool from sglang.srt.model_executor.forward_batch_info import ForwardBatch, ForwardMode from sglang.srt.runtime_context import ( get_parallel, + get_platform, get_spec, ) from sglang.srt.speculative.eagle_utils import per_step_draft_out_cache_loc @@ -81,8 +82,7 @@ from sglang.srt.speculative.ragged_verify import ( read_ragged_verify_mode, resolve_ragged_verify_layout, ) -from sglang.srt.utils import ceil_align, is_cuda, is_sm90_supported, is_xpu -from sglang.srt.utils.common import is_sm120_supported +from sglang.srt.utils import ceil_align, is_cuda, is_xpu if TYPE_CHECKING: from sgl_kernel.flash_mla import FlashMLASchedMeta @@ -91,7 +91,6 @@ if TYPE_CHECKING: from sglang.srt.model_executor.model_runner import ModelRunner from sglang.srt.speculative.ragged_verify import RaggedVerifyLayout -_is_sm120 = is_sm120_supported() _is_cuda = is_cuda() _is_xpu = is_xpu() @@ -147,7 +146,7 @@ def _pad_last_dim(x: T, multiples_of: int = PAGE_INDEX_ALIGNED_SIZE) -> T: def _create_flashmla_metadata(): - if _is_sm120 or _is_xpu: + if get_platform().is_sm120 or _is_xpu: return None import sgl_kernel.flash_mla as flash_mla @@ -574,7 +573,7 @@ class DeepseekV4AttnBackend( model_runner.server_args, "dsv4_prefill_backend", "auto" ) if use_dsv4_q8kv8_sparse_prefill(self.dsv4_prefill_backend): - if not is_sm90_supported(): + if not get_platform().is_sm90: raise ValueError( "DeepSeek-V4 flashmla_sparse_q8 prefill requires SM90 CUDA GPUs." ) @@ -708,7 +707,7 @@ class DeepseekV4AttnBackend( # The SM120 FP4 kernel schedules split_kv=128, while the generic # JIT metadata planner encodes split_kv=256. force_deep_gemm_metadata=( - self.enable_deepseek_v4_fp4_indexer and _is_sm120 + self.enable_deepseek_v4_fp4_indexer and get_platform().is_sm120 ), use_prefill_cuda_graph=use_prefill_cuda_graph, ) @@ -1380,7 +1379,7 @@ class DeepseekV4AttnBackend( return assert isinstance(metadata, DSV4Metadata) - use_sparse_prefill = not _is_sm120 and ( + use_sparse_prefill = not get_platform().is_sm120 and ( num_qo_tokens > _LARGE_INDEXER_QUERY_THRESHOLD or envs.SGLANG_OPT_FLASHMLA_SPARSE_PREFILL.get() ) @@ -1761,7 +1760,7 @@ class DeepseekV4AttnBackend( # sparse_prefill_fwd does not support SM120. if ( forward_batch.forward_mode.is_extend_without_speculative() - and not _is_sm120 + and not get_platform().is_sm120 and ( q.shape[0] > _LARGE_INDEXER_QUERY_THRESHOLD or envs.SGLANG_OPT_FLASHMLA_SPARSE_PREFILL.get() @@ -1787,7 +1786,7 @@ class DeepseekV4AttnBackend( attn_sink=attn_sink, ) - if _is_sm120: + if get_platform().is_sm120: from sglang.kernels.ops.attention.flash_mla_sm120 import ( flash_mla_with_kvcache_sm120, ) diff --git a/python/sglang/srt/layers/attention/dsa/paged_mqa_logits_backend.py b/python/sglang/srt/layers/attention/dsa/paged_mqa_logits_backend.py index 8f46d16fc..d53239bb1 100644 --- a/python/sglang/srt/layers/attention/dsa/paged_mqa_logits_backend.py +++ b/python/sglang/srt/layers/attention/dsa/paged_mqa_logits_backend.py @@ -2,7 +2,8 @@ from __future__ import annotations from enum import Enum -from sglang.srt.utils import is_hip, is_sm100_supported +from sglang.srt.runtime_context import get_platform +from sglang.srt.utils import is_hip class DSAPagedMQALogitsBackend(Enum): @@ -34,7 +35,7 @@ class DSAPagedMQALogitsBackend(Enum): if value == "aiter": raise ValueError("dsa_paged_mqa_logits_backend='aiter' requires ROCm.") if value == "cutedsl": - if not is_sm100_supported(): + if not get_platform().is_sm100: raise ValueError( "dsa_paged_mqa_logits_backend='cutedsl' requires SM100 (Blackwell)." ) diff --git a/python/sglang/srt/layers/attention/dsa_backend.py b/python/sglang/srt/layers/attention/dsa_backend.py index 8f9929a95..9369ad289 100644 --- a/python/sglang/srt/layers/attention/dsa_backend.py +++ b/python/sglang/srt/layers/attention/dsa_backend.py @@ -16,8 +16,10 @@ import torch from sglang.srt.configs.model_config import get_dsa_index_topk, is_deepseek_dsa from sglang.srt.runtime_context import ( + get_buffer, get_exec, get_parallel, + get_platform, get_spec, ) @@ -73,13 +75,11 @@ from sglang.srt.layers.utils.cp_utils import ( cp_split_and_rebuild_position, ) from sglang.srt.model_executor.forward_batch_info import ForwardBatch, ForwardMode -from sglang.srt.runtime_context import get_buffer from sglang.srt.utils import ( get_bool_env_var, is_cuda, is_gfx95_supported, is_hip, - is_sm100_supported, print_warning_once, ) @@ -670,7 +670,7 @@ class DeepseekSparseAttnBackend( forward_mode.is_target_verify() and next_n and next_n >= 2 - and is_sm100_supported() + and get_platform().is_sm100 ): return cache_seqlens_int32.view(-1, 1).expand(-1, next_n).contiguous() if forward_mode.is_target_verify() or forward_mode.is_draft_extend_v2(): @@ -1497,7 +1497,7 @@ class DeepseekSparseAttnBackend( paged_mqa_ctx_lens_2d = None if ( self.speculative_num_draft_tokens >= 2 - and is_sm100_supported() + and get_platform().is_sm100 and metadata.paged_mqa_ctx_lens_2d is not None and metadata.paged_mqa_ctx_lens_2d.dim() == 2 and metadata.paged_mqa_ctx_lens_2d.size(0) == bs diff --git a/python/sglang/srt/layers/attention/dsv4/indexer.py b/python/sglang/srt/layers/attention/dsv4/indexer.py index 774b2729b..d8cbcfa36 100644 --- a/python/sglang/srt/layers/attention/dsv4/indexer.py +++ b/python/sglang/srt/layers/attention/dsv4/indexer.py @@ -40,10 +40,13 @@ from sglang.srt.model_executor.runner_backend_utils.breakable_cuda_graph.context from sglang.srt.model_executor.runner_backend_utils.tc_piecewise_cuda_graph import ( is_in_tc_piecewise_cuda_graph, ) -from sglang.srt.runtime_context import get_exec, get_parallel +from sglang.srt.runtime_context import ( + get_exec, + get_parallel, + get_platform, +) from sglang.srt.state_capturer.indexer_topk import get_global_indexer_capturer from sglang.srt.utils import add_prefix, is_cuda, is_hip, is_xpu -from sglang.srt.utils.common import is_sm120_supported if TYPE_CHECKING: from sglang.srt.layers.attention.base_attn_backend import AttentionBackend @@ -714,7 +717,7 @@ class C4IndexerBackendMixin: elif envs.SGLANG_OPT_USE_AITER_INDEXER.get(): fn = _aiter_fp8_paged_mqa_logits elif envs.SGLANG_FP8_PAGED_MQA_LOGITS_TORCH.get(): - if is_sm120_supported(): + if get_platform().is_sm120: fn = fp8_paged_mqa_logits_torch_sm120 else: fn = fp8_paged_mqa_logits_torch diff --git a/python/sglang/srt/layers/attention/flashinfer_backend.py b/python/sglang/srt/layers/attention/flashinfer_backend.py index 78c00b847..80ec2bb22 100644 --- a/python/sglang/srt/layers/attention/flashinfer_backend.py +++ b/python/sglang/srt/layers/attention/flashinfer_backend.py @@ -1,8 +1,10 @@ from __future__ import annotations from sglang.srt.runtime_context import ( + get_buffer, get_exec, get_parallel, + get_platform, ) """ @@ -44,7 +46,6 @@ from sglang.srt.model_executor.forward_batch_info import ForwardBatch, ForwardMo from sglang.srt.model_executor.runner_backend_utils.tc_piecewise_cuda_graph import ( is_in_tc_piecewise_cuda_graph, ) -from sglang.srt.runtime_context import get_buffer from sglang.srt.speculative.spec_info import SpecInput, SpecInputType from sglang.srt.speculative.spec_utils import ( draft_kv_indices_buffer_width, @@ -55,7 +56,6 @@ from sglang.srt.utils import ( get_cuda_graph_max_batch_size, get_int_env_var, is_flashinfer_available, - is_sm100_supported, next_power_of_2, ) @@ -471,7 +471,7 @@ class FlashInferAttnBackend(AttentionBackend): ] fmha_backend = "auto" - if is_sm100_supported(): + if get_platform().is_sm100: # Disable CUTLASS backend when piecewise cuda graph is enabled # due to TMA descriptor initialization issues on SM100 GPUs. if not check_cuda_graph_backend(Phase.PREFILL, Backend.TC_PIECEWISE): diff --git a/python/sglang/srt/layers/attention/flashinfer_mla_backend.py b/python/sglang/srt/layers/attention/flashinfer_mla_backend.py index 49cf04220..eed97d760 100644 --- a/python/sglang/srt/layers/attention/flashinfer_mla_backend.py +++ b/python/sglang/srt/layers/attention/flashinfer_mla_backend.py @@ -1,9 +1,11 @@ from __future__ import annotations from sglang.srt.runtime_context import ( + get_buffer, get_disagg, get_exec, get_parallel, + get_platform, get_schedule, ) @@ -41,7 +43,6 @@ from sglang.srt.model_executor.runner_backend_utils.breakable_cuda_graph import from sglang.srt.model_executor.runner_backend_utils.tc_piecewise_cuda_graph import ( is_in_tc_piecewise_cuda_graph, ) -from sglang.srt.runtime_context import get_buffer from sglang.srt.speculative.spec_info import SpecInput, SpecInputType from sglang.srt.speculative.spec_utils import ( draft_kv_indices_buffer_width, @@ -50,7 +51,6 @@ from sglang.srt.speculative.spec_utils import ( ) from sglang.srt.utils import ( is_flashinfer_available, - is_sm100_supported, next_power_of_2, ) @@ -277,7 +277,7 @@ class FlashInferMLAAttnBackend(AttentionBackend): else: self.q_indptr_decode = q_indptr_decode_buf - if is_sm100_supported(): + if get_platform().is_sm100: self.fmha_backend = "cutlass" else: self.fmha_backend = "auto" diff --git a/python/sglang/srt/layers/attention/minicpm/backend.py b/python/sglang/srt/layers/attention/minicpm/backend.py index 45e181c48..126c9e859 100644 --- a/python/sglang/srt/layers/attention/minicpm/backend.py +++ b/python/sglang/srt/layers/attention/minicpm/backend.py @@ -17,8 +17,12 @@ from sglang.srt.layers.attention.minicpm.attention_adapter import ( ) from sglang.srt.layers.attention.minicpm.cache import attach_compressed_cache from sglang.srt.model_executor.forward_batch_info import ForwardBatch -from sglang.srt.runtime_context import get_parallel, get_schedule -from sglang.srt.utils import is_blackwell_supported, next_power_of_2 +from sglang.srt.runtime_context import ( + get_parallel, + get_platform, + get_schedule, +) +from sglang.srt.utils import next_power_of_2 if TYPE_CHECKING: from sglang.srt.layers.radix_attention import RadixAttention @@ -127,7 +131,7 @@ class MiniCPMSparseBackend(AttentionBackend): use_flashinfer: bool, ): super().__init__() - use_blackwell = is_blackwell_supported() + use_blackwell = get_platform().is_blackwell if use_blackwell: fa_impl_ver = 4 self.flash_attn_backend = FlashAttentionBackend( diff --git a/python/sglang/srt/layers/attention/minimax_sparse_backend.py b/python/sglang/srt/layers/attention/minimax_sparse_backend.py index 55808c74f..814a70d79 100644 --- a/python/sglang/srt/layers/attention/minimax_sparse_backend.py +++ b/python/sglang/srt/layers/attention/minimax_sparse_backend.py @@ -21,6 +21,10 @@ from sglang.srt.layers.attention.base_attn_backend import ( ) from sglang.srt.mem_cache.memory_pool import MiniMaxSparseKVPool from sglang.srt.model_executor.forward_batch_info import ForwardBatch, ForwardMode +from sglang.srt.runtime_context import ( + get_parallel, + get_spec, +) from sglang.srt.server_args import m3_fp8_attn_gemm_enabled from sglang.srt.utils import is_npu @@ -222,7 +226,6 @@ class MiniMaxSparseAttnBackend(AttentionBackend): self._msa_dec_meta = None if self.use_msa: - from sglang.srt.runtime_context import get_parallel self.num_q_heads = ( runner.model_config.num_attention_heads // get_parallel().attn_tp_size @@ -247,7 +250,6 @@ class MiniMaxSparseAttnBackend(AttentionBackend): Phase, check_cuda_graph_backend, ) - from sglang.srt.runtime_context import get_spec spec = get_spec() self.speculative_num_draft_tokens = spec.speculative_num_draft_tokens diff --git a/python/sglang/srt/layers/attention/tokenspeed_mla_backend.py b/python/sglang/srt/layers/attention/tokenspeed_mla_backend.py index 7000c0fa0..91ed1bd50 100644 --- a/python/sglang/srt/layers/attention/tokenspeed_mla_backend.py +++ b/python/sglang/srt/layers/attention/tokenspeed_mla_backend.py @@ -56,7 +56,11 @@ from sglang.srt.layers.attention.trtllm_mla_backend import ( ) from sglang.srt.layers.dcp.layout import get_dcp_lens from sglang.srt.layers.logits_processor import get_in_autotune_dummy_run -from sglang.srt.runtime_context import get_parallel, max_speculative_num_draft_tokens +from sglang.srt.runtime_context import ( + get_parallel, + get_resources, + max_speculative_num_draft_tokens, +) from sglang.srt.utils import is_flashinfer_available, is_tokenspeed_mla_available if is_flashinfer_available(): @@ -87,7 +91,6 @@ def _get_tokenspeed_workspace( kv_lora_rank: int, max_q_len: int = _TOKENSPEED_MAX_Q_LEN, ) -> torch.Tensor: - from sglang.srt.runtime_context import get_resources # DCP target verification gathers Q to the full head count before launching # TokenSpeed; size for that launch shape, not the rank-local head count. diff --git a/python/sglang/srt/layers/attention/trtllm_mha_backend.py b/python/sglang/srt/layers/attention/trtllm_mha_backend.py index dfdebf508..e60083627 100644 --- a/python/sglang/srt/layers/attention/trtllm_mha_backend.py +++ b/python/sglang/srt/layers/attention/trtllm_mha_backend.py @@ -45,13 +45,17 @@ from sglang.srt.layers.radix_attention import AttentionType from sglang.srt.mem_cache.memory_pool import KVWriteLoc from sglang.srt.mem_cache.swa_memory_pool import SWAKVPool from sglang.srt.model_executor.forward_batch_info import ForwardBatch, ForwardMode -from sglang.srt.runtime_context import get_buffer, get_spec +from sglang.srt.runtime_context import ( + get_buffer, + get_parallel, + get_platform, + get_spec, +) from sglang.srt.speculative.ragged_verify import ( build_ragged_target_verify_geometry, resolve_ragged_verify_layout, ) from sglang.srt.utils import is_flashinfer_available -from sglang.srt.utils.common import is_sm90_supported, is_sm120_supported logger = logging.getLogger(__name__) @@ -222,10 +226,10 @@ class TRTLLMHAAttnBackend(FlashInferAttnBackend): # TRTLLM-GEN: # KV bf16: q_type = bf16, out_type=model_runner.dtype # KV fp8: q_type = fp8, out_type=model_runner.dtype - self.is_xqa_impl = is_sm90_supported() or is_sm120_supported() + self.is_xqa_impl = get_platform().is_sm90 or get_platform().is_sm120 # fmha_v2 prefill kernel supports SM90 and SM120 - self.use_fmha_v2 = is_sm90_supported() or is_sm120_supported() + self.use_fmha_v2 = get_platform().is_sm90 or get_platform().is_sm120 # trtllm-gen serves page_size >= 128 only through its dynamic # tokens-per-page kernels, which exist solely for GQA with equal QK/V @@ -234,7 +238,6 @@ class TRTLLMHAAttnBackend(FlashInferAttnBackend): # "Missing TRTLLM-GEN kernel" error during CUDA-graph capture. # XQA (SM90/SM120 decode) has native page-128 kernels; no check needed. if self.page_size >= 128 and not self.is_xqa_impl: - from sglang.srt.runtime_context import get_parallel attn_tp_size = get_parallel().attn_tp_size num_q_heads = config.num_attention_heads // attn_tp_size diff --git a/python/sglang/srt/layers/attention/vision.py b/python/sglang/srt/layers/attention/vision.py index 20e87619d..4661fbd89 100644 --- a/python/sglang/srt/layers/attention/vision.py +++ b/python/sglang/srt/layers/attention/vision.py @@ -17,12 +17,17 @@ from sglang.kernels.ops.layernorm.norm import ( ) from sglang.srt.environ import envs from sglang.srt.models.utils import apply_qk_norm -from sglang.srt.runtime_context import get_context, get_exec, get_mm, get_parallel +from sglang.srt.runtime_context import ( + get_context, + get_exec, + get_mm, + get_parallel, + get_platform, +) from sglang.srt.utils import ( cpu_has_amx_support, get_bool_env_var, get_device_capability, - is_blackwell_supported, is_cpu, is_cuda, is_hip, @@ -1268,7 +1273,7 @@ class VisionAttention(nn.Module): backend = "xpu_attn" else: backend = "sdpa" - if backend == "fa3" and is_blackwell_supported(): + if backend == "fa3" and get_platform().is_blackwell: raise ValueError("The 'fa3' backend is not supported on Blackwell GPUs") return backend diff --git a/python/sglang/srt/layers/communicator.py b/python/sglang/srt/layers/communicator.py index e92806fd4..0c1dd372f 100644 --- a/python/sglang/srt/layers/communicator.py +++ b/python/sglang/srt/layers/communicator.py @@ -73,7 +73,13 @@ from sglang.srt.model_executor.cuda_graph_config import ( check_cuda_graph_backend, ) from sglang.srt.model_executor.forward_batch_info import ForwardBatch -from sglang.srt.runtime_context import get_exec, get_forward, get_parallel, get_spec +from sglang.srt.runtime_context import ( + get_exec, + get_forward, + get_parallel, + get_platform, + get_spec, +) from sglang.srt.speculative.spec_info import SpeculativeAlgorithm from sglang.srt.utils import ( get_bool_env_var, @@ -82,14 +88,12 @@ from sglang.srt.utils import ( is_gfx95_supported, is_hip, is_npu, - is_sm90_supported, - is_sm100_supported, ) _is_cuda = is_cuda() _is_flashinfer_available = is_flashinfer_available() -_is_sm90_supported = _is_cuda and is_sm90_supported() -_is_sm100_supported = _is_cuda and is_sm100_supported() +_is_sm90_supported = _is_cuda and get_platform().is_sm90 +_is_sm100_supported = _is_cuda and get_platform().is_sm100 _use_aiter = get_bool_env_var("SGLANG_USE_AITER") and is_hip() _is_gfx95_supported = is_gfx95_supported() _is_npu = is_npu() diff --git a/python/sglang/srt/layers/cp/bcg.py b/python/sglang/srt/layers/cp/bcg.py index 6c26099dd..9be7d7111 100644 --- a/python/sglang/srt/layers/cp/bcg.py +++ b/python/sglang/srt/layers/cp/bcg.py @@ -22,6 +22,7 @@ from typing import TYPE_CHECKING, Any, Dict, Optional import torch from sglang.srt.arg_groups.overrides import ( + attention_backends_of, resolved_view, resolving_view, ) @@ -46,7 +47,6 @@ if TYPE_CHECKING: def supports_prefill_cp_bcg(server_args: ServerArgs) -> bool: """Return whether the selected prefill-CP configuration supports BCG.""" - from sglang.srt.arg_groups.overrides import attention_backends_of cfg = resolving_view(server_args) resolved = resolved_view(server_args) diff --git a/python/sglang/srt/layers/deep_gemm_wrapper/compile_utils.py b/python/sglang/srt/layers/deep_gemm_wrapper/compile_utils.py index fe2b659f4..eb3e5988d 100644 --- a/python/sglang/srt/layers/deep_gemm_wrapper/compile_utils.py +++ b/python/sglang/srt/layers/deep_gemm_wrapper/compile_utils.py @@ -19,11 +19,11 @@ from sglang.srt.environ import envs from sglang.srt.layers.deep_gemm_wrapper.configurer import ENABLE_JIT_DEEPGEMM from sglang.srt.model_executor.forward_batch_info import ForwardMode from sglang.srt.runtime_context import ( + get_device, get_disagg, get_parallel, get_schedule, ) -from sglang.srt.server_args import ServerArgs from sglang.srt.utils import ceil_align, ceil_div, get_available_gpu_memory, is_musa logger = logging.getLogger(__name__) @@ -51,7 +51,7 @@ os.environ["DG_JIT_CACHE_DIR"] = envs.SGLANG_DG_CACHE_DIR.get() os.environ["DG_JIT_USE_NVRTC"] = os.getenv("SGL_DG_USE_NVRTC", "0") -def update_deep_gemm_config(gpu_id: int, server_args: ServerArgs): +def update_deep_gemm_config(gpu_id: int): global _BUILTIN_M_LIST global _DO_COMPILE_ALL global _IS_FIRST_RANK_ON_NODE @@ -96,7 +96,7 @@ def update_deep_gemm_config(gpu_id: int, server_args: ServerArgs): m_max = min(1024 * 128, m_max) _BUILTIN_M_LIST += list(range(1, m_max + 1)) - _IS_FIRST_RANK_ON_NODE = server_args.base_gpu_id == gpu_id + _IS_FIRST_RANK_ON_NODE = get_device().base_gpu_id == gpu_id # Check if is the first rank on node. # Default each rank will try compile all Ms to diff --git a/python/sglang/srt/layers/deep_gemm_wrapper/configurer.py b/python/sglang/srt/layers/deep_gemm_wrapper/configurer.py index 09e14f69d..17f04b2a6 100644 --- a/python/sglang/srt/layers/deep_gemm_wrapper/configurer.py +++ b/python/sglang/srt/layers/deep_gemm_wrapper/configurer.py @@ -1,11 +1,11 @@ import logging from sglang.srt.environ import envs +from sglang.srt.runtime_context import get_platform from sglang.srt.utils import ( get_device_sm, is_cuda, is_musa, - is_sm100_supported, ) logger = logging.getLogger(__name__) @@ -34,6 +34,6 @@ def _compute_enable_deep_gemm(): ENABLE_JIT_DEEPGEMM = _compute_enable_deep_gemm() -DEEPGEMM_BLACKWELL = ENABLE_JIT_DEEPGEMM and is_sm100_supported() +DEEPGEMM_BLACKWELL = ENABLE_JIT_DEEPGEMM and get_platform().is_sm100 DEEPGEMM_SCALE_UE8M0 = DEEPGEMM_BLACKWELL DEEPGEMM_NEED_TMA_ALIGNED_SCALES = not (DEEPGEMM_SCALE_UE8M0 or _is_musa) diff --git a/python/sglang/srt/layers/deep_gemm_wrapper/entrypoint.py b/python/sglang/srt/layers/deep_gemm_wrapper/entrypoint.py index 6151b3cd7..fce7d735a 100644 --- a/python/sglang/srt/layers/deep_gemm_wrapper/entrypoint.py +++ b/python/sglang/srt/layers/deep_gemm_wrapper/entrypoint.py @@ -12,7 +12,6 @@ from sglang.srt.layers.deep_gemm_wrapper.configurer import ( # noqa: F401 DEEPGEMM_SCALE_UE8M0, ENABLE_JIT_DEEPGEMM, ) -from sglang.srt.server_args import ServerArgs logger = logging.getLogger(__name__) @@ -245,13 +244,13 @@ def tf32_hc_prenorm_gemm( deep_gemm.tf32_hc_prenorm_gemm(x, fn, out, sqrsum, num_splits=num_splits) -def update_deep_gemm_config(gpu_id: int, server_args: ServerArgs): +def update_deep_gemm_config(gpu_id: int): # deep_gemm.set_pdl can initialize CUDA state, so run it only after the # scheduler/TP worker has been forked and assigned a GPU. if envs.SGLANG_DEEPGEMM_PDL.get() and hasattr(deep_gemm, "set_pdl"): deep_gemm.set_pdl(True) - compile_utils.update_deep_gemm_config(gpu_id, server_args) + compile_utils.update_deep_gemm_config(gpu_id) @contextmanager diff --git a/python/sglang/srt/layers/dp_attention.py b/python/sglang/srt/layers/dp_attention.py index 618550e6b..87117547a 100644 --- a/python/sglang/srt/layers/dp_attention.py +++ b/python/sglang/srt/layers/dp_attention.py @@ -34,7 +34,10 @@ from sglang.srt.runtime_context import ( get_device, get_exec, get_flags, + get_forward, get_parallel, + get_resources, + get_stream, ) from sglang.srt.utils import get_bool_env_var, is_cpu, is_hip @@ -166,7 +169,6 @@ class _DpGatheredBufferWrapper: @classmethod def set_metadata(cls, hidden_size: int, dtype: torch.dtype, device: torch.device): - from sglang.srt.runtime_context import get_flags dp = get_flags().dp dp.buffer_hidden_size = hidden_size @@ -190,7 +192,6 @@ class _DpGatheredBufferWrapper: @classmethod def get_global_dp_buffer(cls, group: GroupCoordinator) -> torch.Tensor: - from sglang.srt.runtime_context import get_flags dp = get_flags().dp with use_symmetric_memory(group, disabled=not cls._dp_max_padding): @@ -203,7 +204,6 @@ class _DpGatheredBufferWrapper: @classmethod def get_local_dp_buffer(cls, group: GroupCoordinator) -> torch.Tensor: - from sglang.srt.runtime_context import get_flags dp = get_flags().dp with use_symmetric_memory(group, disabled=not cls._dp_max_padding): @@ -236,19 +236,16 @@ class _DpGatheredBufferWrapper: @classmethod def get_dp_hidden_size(cls) -> int: - from sglang.srt.runtime_context import get_flags return get_flags().dp.buffer_hidden_size @classmethod def get_dp_dtype(cls) -> torch.dtype: - from sglang.srt.runtime_context import get_flags return get_flags().dp.buffer_dtype @classmethod def get_dp_device(cls) -> torch.device: - from sglang.srt.runtime_context import get_flags return get_flags().dp.buffer_device @@ -313,13 +310,11 @@ def set_is_extend_in_batch(is_extend_in_batch: bool): # Sticky within the thread: every ForwardBatch construction writes it, # graph runners force False around capture; readers are the EP # dispatchers on the same (single) forward thread. - from sglang.srt.runtime_context import get_forward get_forward().set("is_extend_in_batch", is_extend_in_batch) def get_is_extend_in_batch() -> bool: - from sglang.srt.runtime_context import get_forward return get_forward().is_extend_in_batch @@ -909,7 +904,6 @@ def dp_reduce_scatter_tensor(output: torch.Tensor, input: torch.Tensor): # deadlock on the RCCL communicator), each overlapping the other's compute. # --------------------------------------------------------------------------- def get_dp_tbo_comm_stream() -> torch.cuda.Stream: - from sglang.srt.runtime_context import get_stream return get_stream("dp_tbo_comm") @@ -921,7 +915,6 @@ def get_dp_tbo_comm_stream() -> torch.cuda.Stream: # ("...create internal OS-specific events"). Reuse one event per (kind, subbatch) # and just re-record it (mirrors the mori CommStreamPool event reuse). def _tbo_event(key) -> torch.cuda.Event: - from sglang.srt.runtime_context import get_resources pool = get_resources().tbo_event_pool ev = pool.get(key) diff --git a/python/sglang/srt/layers/flashinfer_comm_fusion.py b/python/sglang/srt/layers/flashinfer_comm_fusion.py index 5715fce7b..00dab3996 100644 --- a/python/sglang/srt/layers/flashinfer_comm_fusion.py +++ b/python/sglang/srt/layers/flashinfer_comm_fusion.py @@ -13,13 +13,16 @@ from sglang.srt.distributed import ( get_tp_group, ) from sglang.srt.distributed.parallel_state import in_the_same_node_as -from sglang.srt.runtime_context import get_exec, get_parallel +from sglang.srt.runtime_context import ( + get_exec, + get_parallel, + get_platform, + get_resources, +) from sglang.srt.utils import ( ceil_align, get_cuda_driver_bindings, is_flashinfer_available, - is_sm90_supported, - is_sm100_supported, ) from sglang.srt.utils.custom_op import register_custom_op @@ -38,27 +41,27 @@ _flashinfer_allreduce_supports_trigger_completion = False def _mnnvl_supported(is_multi_node: bool) -> bool: """Whether the mnnvl backend is usable on the current system.""" - if is_sm100_supported(): + if get_platform().is_sm100: return True - return is_sm90_supported() and not is_multi_node + return get_platform().is_sm90 and not is_multi_node def _resolve_backend(backend: str, is_multi_node: bool = False) -> str: """Resolve the requested FlashInfer allreduce fusion backend.""" - if not (is_sm90_supported() or is_sm100_supported()): + if not (get_platform().is_sm90 or get_platform().is_sm100): raise ValueError( "FlashInfer allreduce fusion requires SM90 or SM10X NVIDIA GPUs." ) if backend == "auto": if is_multi_node: - if is_sm100_supported(): + if get_platform().is_sm100: return "mnnvl" raise ValueError( "FlashInfer allreduce fusion does not support multi-node on " "non-Blackwell systems." ) - if is_sm100_supported(): + if get_platform().is_sm100: return "mnnvl" return "trtllm" @@ -630,7 +633,6 @@ class FlashInferWorkspaceManager: def _get_workspace_manager(use_attn_tp_group: bool) -> FlashInferWorkspaceManager: """The per-group fusion workspace manager; the instances live on ``ctx.resources`` (one per comm group, created lazily).""" - from sglang.srt.runtime_context import get_resources buffers = get_resources().buffers name = ( @@ -1018,7 +1020,6 @@ def pre_initialize_workspaces( def cleanup_flashinfer_workspace(): - from sglang.srt.runtime_context import get_resources buffers = get_resources().buffers for name in ( diff --git a/python/sglang/srt/layers/flashinfer_mnnvl_cutedsl.py b/python/sglang/srt/layers/flashinfer_mnnvl_cutedsl.py index 6b91d9e81..653197c84 100644 --- a/python/sglang/srt/layers/flashinfer_mnnvl_cutedsl.py +++ b/python/sglang/srt/layers/flashinfer_mnnvl_cutedsl.py @@ -12,6 +12,8 @@ from typing import TYPE_CHECKING import torch import torch.distributed as dist +from sglang.srt.runtime_context import get_spec + if TYPE_CHECKING: from torch.distributed import ProcessGroup @@ -149,7 +151,6 @@ class FlashInferMNNVLCuteDSLARFusion: ) = _import_kernel_backend() # Only fused finalize launches have a completed shared-expert handoff; # standalone AllReduce kernels retain the safe load ordering. - from sglang.srt.runtime_context import get_spec if get_spec().speculative_algorithm is None: self.workspace_config = _with_early_finalize_shared_load(default_config) diff --git a/python/sglang/srt/layers/k3_ar_fusion.py b/python/sglang/srt/layers/k3_ar_fusion.py index e2f72c008..b91774a5c 100644 --- a/python/sglang/srt/layers/k3_ar_fusion.py +++ b/python/sglang/srt/layers/k3_ar_fusion.py @@ -25,6 +25,7 @@ import torch import sglang.srt.runtime_context as ctx from sglang.kernels.jit.utils import cache_once from sglang.srt.environ import envs +from sglang.srt.runtime_context import get_parallel if TYPE_CHECKING: from sglang.srt.distributed.device_communicators.custom_all_reduce_v2 import ( @@ -84,7 +85,6 @@ def _get_state() -> Optional[_State]: CustomAllReduceV2, ) from sglang.srt.distributed.parallel_state import get_tp_group - from sglang.srt.runtime_context import get_parallel if get_parallel().tp_size <= 1: return None diff --git a/python/sglang/srt/layers/k3_gemm_ar.py b/python/sglang/srt/layers/k3_gemm_ar.py index ff902f66d..9a36809a8 100644 --- a/python/sglang/srt/layers/k3_gemm_ar.py +++ b/python/sglang/srt/layers/k3_gemm_ar.py @@ -16,6 +16,7 @@ from typing import TYPE_CHECKING import torch from sglang.srt.environ import envs +from sglang.srt.runtime_context import get_parallel if TYPE_CHECKING: from sglang.srt.layers.linear import RowParallelLinear @@ -33,7 +34,6 @@ def _init() -> bool: _INITIALIZED = True if not envs.SGLANG_K3_GEMM_AR.get(): return False - from sglang.srt.runtime_context import get_parallel world_size = get_parallel().tp_size if not (2 <= world_size <= 8): @@ -57,7 +57,6 @@ def maybe_wrap_o_proj(o_proj: RowParallelLinear) -> None: return from sglang.kernels.ops.kimi_k3 import gemm_ar as mod from sglang.srt.distributed.parallel_state import get_tp_group - from sglang.srt.runtime_context import get_parallel parallel = get_parallel() world_size = parallel.tp_size diff --git a/python/sglang/srt/layers/k3_sp_collective.py b/python/sglang/srt/layers/k3_sp_collective.py index b63c75b8c..9ab74858a 100644 --- a/python/sglang/srt/layers/k3_sp_collective.py +++ b/python/sglang/srt/layers/k3_sp_collective.py @@ -17,6 +17,7 @@ import torch from sglang.srt.environ import envs from sglang.srt.layers import k3_ar_fusion +from sglang.srt.runtime_context import get_exec, get_parallel if TYPE_CHECKING: from sglang.srt.distributed.device_communicators.custom_all_reduce_v2 import ( @@ -64,7 +65,6 @@ def _init_state() -> Optional[_State]: from sglang.srt.distributed.device_communicators.custom_all_reduce_v2 import ( CustomAllReduceV2, ) - from sglang.srt.runtime_context import get_exec, get_parallel from sglang.srt.utils.common import get_device_sm a2a = get_exec().moe.moe_a2a_backend diff --git a/python/sglang/srt/layers/moe/cutlass_moe.py b/python/sglang/srt/layers/moe/cutlass_moe.py index 59ddf6f40..ebc738c70 100755 --- a/python/sglang/srt/layers/moe/cutlass_moe.py +++ b/python/sglang/srt/layers/moe/cutlass_moe.py @@ -4,7 +4,8 @@ from typing import Optional, Tuple import torch -from sglang.srt.utils import is_cuda, is_sm90_supported, is_sm100_supported +from sglang.srt.runtime_context import get_platform +from sglang.srt.utils import is_cuda _is_cuda = is_cuda() if _is_cuda: @@ -145,7 +146,7 @@ def cutlass_fused_experts_fp8( if use_mxfp8: assert es_up and es_down, "MXFP8 requires expert-specialization for both GEMMs" - assert is_sm100_supported(), "MXFP8 requires SM100" + assert get_platform().is_sm100, "MXFP8 requires SM100" assert k % 32 == 0, "MXFP8 requires hidden size to be divisible by 32" assert n % 32 == 0, "MXFP8 requires intermediate size to be divisible by 32" assert w1_scale.dtype == torch.uint8, "MXFP8 w1_scale must be uint8" @@ -222,7 +223,7 @@ def cutlass_fused_experts_fp8( a_sf_layout = torch.empty((num_experts, 5), device=device, dtype=torch.int) w_sf_layout = torch.empty((num_experts, 5), device=device, dtype=torch.int) - if is_sm90_supported() and es_up: + if get_platform().is_sm90 and es_up: es_fp8_blockwise_scaled_grouped_mm( c1, rep_a_q, @@ -288,7 +289,7 @@ def cutlass_fused_experts_fp8( else: intemediate_q, a2_scale = sglang_per_token_group_quant_fp8(intermediate, 128) - if is_sm90_supported() and es_down: + if get_platform().is_sm90 and es_down: es_fp8_blockwise_scaled_grouped_mm( c2, intemediate_q, diff --git a/python/sglang/srt/layers/moe/fused_moe_triton/triton_kernels_moe.py b/python/sglang/srt/layers/moe/fused_moe_triton/triton_kernels_moe.py index 199dfe3ae..d410fde58 100644 --- a/python/sglang/srt/layers/moe/fused_moe_triton/triton_kernels_moe.py +++ b/python/sglang/srt/layers/moe/fused_moe_triton/triton_kernels_moe.py @@ -19,10 +19,10 @@ from triton_kernels.numerics import InFlexData from triton_kernels.swiglu import swiglu_fn from triton_kernels.tensor import FP4 +from sglang.srt.runtime_context import get_platform from sglang.srt.utils import is_cuda -from sglang.srt.utils.common import is_sm120_supported -if is_sm120_supported(): +if get_platform().is_sm120: # use the regular gather/scatter implementation for unsupported devices. update_opt_flags_constraints({"is_persistent": False}) diff --git a/python/sglang/srt/layers/moe/moe_runner/aiter.py b/python/sglang/srt/layers/moe/moe_runner/aiter.py index 8737b93c2..66311cc4f 100644 --- a/python/sglang/srt/layers/moe/moe_runner/aiter.py +++ b/python/sglang/srt/layers/moe/moe_runner/aiter.py @@ -19,6 +19,7 @@ from sglang.srt.layers.moe.moe_runner.base import ( register_pre_permute, ) from sglang.srt.layers.moe.utils import MoeRunnerBackend +from sglang.srt.runtime_context import get_parallel from sglang.srt.utils import get_bool_env_var, get_int_env_var if TYPE_CHECKING: @@ -172,7 +173,6 @@ def _mori_decode_recv_bound(recv_rows: int, topk: int) -> int: get_dp_global_num_tokens, get_is_extend_in_batch, ) - from sglang.srt.runtime_context import get_parallel if get_is_extend_in_batch(): return 0 diff --git a/python/sglang/srt/layers/moe/moe_runner/base.py b/python/sglang/srt/layers/moe/moe_runner/base.py index d1142a7ae..e5a2b23d9 100644 --- a/python/sglang/srt/layers/moe/moe_runner/base.py +++ b/python/sglang/srt/layers/moe/moe_runner/base.py @@ -12,6 +12,7 @@ from sglang.srt.layers.moe.utils import ( MoeRunnerBackendLike, RoutingMethodType, ) +from sglang.srt.runtime_context import get_forward if TYPE_CHECKING: from sglang.srt.layers.moe.moe_runner.triton import ( @@ -29,7 +30,6 @@ if TYPE_CHECKING: def moe_output_buffer_ctx(buf: torch.Tensor): """Provide the MoE output buffer for the current forward scope.""" - from sglang.srt.runtime_context import get_forward return get_forward().scoped(moe_output_buffer=buf) diff --git a/python/sglang/srt/layers/moe/moe_runner/deep_gemm.py b/python/sglang/srt/layers/moe/moe_runner/deep_gemm.py index 17d075ff2..421b25ab5 100644 --- a/python/sglang/srt/layers/moe/moe_runner/deep_gemm.py +++ b/python/sglang/srt/layers/moe/moe_runner/deep_gemm.py @@ -30,7 +30,11 @@ from sglang.srt.layers.moe.moe_runner.base import ( register_pre_permute, ) from sglang.srt.layers.moe.utils import MoeRunnerBackend, get_moe_a2a_backend -from sglang.srt.runtime_context import get_exec, get_flags +from sglang.srt.runtime_context import ( + get_exec, + get_flags, + get_parallel, +) from sglang.srt.utils import ( ceil_div, dispose_tensor, @@ -1077,7 +1081,6 @@ def pre_permute_flashinfer_to_deep_gemm( """Feed one-sided A2A output into DeepGEMM with fused expert remapping.""" from sglang.srt.layers.moe.token_dispatcher.standard import StandardDispatchOutput - from sglang.srt.runtime_context import get_parallel if dispatch_output.hidden_states.dtype != torch.bfloat16: raise TypeError( diff --git a/python/sglang/srt/layers/moe/token_dispatcher/deepep.py b/python/sglang/srt/layers/moe/token_dispatcher/deepep.py index 2cedb2bcb..584511146 100644 --- a/python/sglang/srt/layers/moe/token_dispatcher/deepep.py +++ b/python/sglang/srt/layers/moe/token_dispatcher/deepep.py @@ -65,6 +65,8 @@ from enum import Enum, IntEnum, auto import torch import torch.distributed as dist +from sglang.srt.runtime_context import get_resources + _use_aiter = get_bool_env_var("SGLANG_USE_AITER") and is_hip() logger = logging.getLogger(__name__) @@ -180,8 +182,6 @@ class DeepEPBuffer: def _state(cls): from types import SimpleNamespace - from sglang.srt.runtime_context import get_resources - buffers = get_resources().buffers state = buffers.get("deepep_ep_state") if state is None: diff --git a/python/sglang/srt/layers/moe/token_dispatcher/deepep_v2.py b/python/sglang/srt/layers/moe/token_dispatcher/deepep_v2.py index 6f56cd305..beb2dc372 100644 --- a/python/sglang/srt/layers/moe/token_dispatcher/deepep_v2.py +++ b/python/sglang/srt/layers/moe/token_dispatcher/deepep_v2.py @@ -21,6 +21,10 @@ from sglang.srt.layers.moe.utils import ( DeepEPv2Fp8ScaleFormat, get_deepep_v2_fp8_scale_format, ) +from sglang.srt.runtime_context import ( + get_exec, + get_resources, +) logger = logging.getLogger(__name__) @@ -112,7 +116,6 @@ def _ensure_fp8_quant_available() -> None: def _get_allow_hybrid_mode() -> bool: - from sglang.srt.runtime_context import get_exec return get_exec().moe.deepep_v2_mode == "hybrid" @@ -139,8 +142,6 @@ class DeepEPv2Buffer: def _state(cls): from types import SimpleNamespace - from sglang.srt.runtime_context import get_resources - buffers = get_resources().buffers state = buffers.get(cls._STATE_KEY) if state is None: diff --git a/python/sglang/srt/layers/moe/token_dispatcher/mooncake.py b/python/sglang/srt/layers/moe/token_dispatcher/mooncake.py index 0a8d9ec55..b059db02c 100644 --- a/python/sglang/srt/layers/moe/token_dispatcher/mooncake.py +++ b/python/sglang/srt/layers/moe/token_dispatcher/mooncake.py @@ -20,6 +20,7 @@ from sglang.srt.layers.moe.token_dispatcher.base import ( ) from sglang.srt.layers.moe.topk import TopKOutput from sglang.srt.layers.moe.utils import DeepEPMode +from sglang.srt.runtime_context import get_resources from sglang.srt.utils import get_int_env_var logger = logging.getLogger(__name__) @@ -64,8 +65,6 @@ class EPBuffer: def _state(cls): from types import SimpleNamespace - from sglang.srt.runtime_context import get_resources - buffers = get_resources().buffers state = buffers.get("mooncake_ep_state") if state is None: diff --git a/python/sglang/srt/layers/moe/token_dispatcher/nixl.py b/python/sglang/srt/layers/moe/token_dispatcher/nixl.py index 2f2e3e849..7b64b3381 100644 --- a/python/sglang/srt/layers/moe/token_dispatcher/nixl.py +++ b/python/sglang/srt/layers/moe/token_dispatcher/nixl.py @@ -23,7 +23,10 @@ from sglang.srt.layers.moe.token_dispatcher.deepep import ( ) from sglang.srt.layers.moe.topk import TopKOutput from sglang.srt.layers.moe.utils import DeepEPMode -from sglang.srt.runtime_context import get_parallel +from sglang.srt.runtime_context import ( + get_parallel, + get_resources, +) try: from nixl_ep import Buffer @@ -53,8 +56,6 @@ class NixlEPBuffer: def _state(cls): from types import SimpleNamespace - from sglang.srt.runtime_context import get_resources - buffers = get_resources().buffers state = buffers.get("nixl_ep_state") if state is None: diff --git a/python/sglang/srt/layers/moe/topk.py b/python/sglang/srt/layers/moe/topk.py index ebb0580ea..e7431a2f2 100644 --- a/python/sglang/srt/layers/moe/topk.py +++ b/python/sglang/srt/layers/moe/topk.py @@ -37,7 +37,12 @@ import triton.language as tl if TYPE_CHECKING: from triton_kernels.tensor_details.ragged_tensor import RaggedTensorMetadata -from sglang.srt.runtime_context import get_exec, get_lora, get_parallel +from sglang.srt.runtime_context import ( + get_exec, + get_lora, + get_parallel, + get_server_args, +) try: from triton_kernels.tensor import make_ragged_tensor_metadata @@ -1523,7 +1528,6 @@ def _eplb_remap_enabled() -> bool: # initial expert placement is non-trivial, or there are redundant physical # experts. Otherwise the map is identity and the remap must be skipped (it is # both unnecessary and not well-defined over the padded region of topk_ids). - from sglang.srt.runtime_context import get_server_args try: get_server_args() # probes that a config is published diff --git a/python/sglang/srt/layers/moe/utils.py b/python/sglang/srt/layers/moe/utils.py index 21291ed56..e2b4635d0 100644 --- a/python/sglang/srt/layers/moe/utils.py +++ b/python/sglang/srt/layers/moe/utils.py @@ -19,13 +19,13 @@ from sglang.srt.runtime_context import ( get_forward, get_model, get_parallel, + get_server_args, get_spec, ) from sglang.srt.utils import is_cuda, is_npu _is_npu = is_npu() -from sglang.srt.runtime_context import get_server_args from sglang.srt.utils.common import log_info_on_rank0 logger = logging.getLogger(__name__) @@ -486,7 +486,6 @@ def is_shared_experts_fusion_disabled() -> bool: ) moe = get_flags().moe if moe.disable_shared_experts_fusion is None: - from sglang.srt.runtime_context import get_exec return get_exec().moe.disable_shared_experts_fusion return moe.disable_shared_experts_fusion @@ -529,7 +528,6 @@ def install_shared_experts_fusion_decision( Inside ``draft_model_build_scope`` the answer also lands on the speculative leaf, so a flags dump afterwards shows both runners' decisions. """ - from sglang.srt.runtime_context import get_exec disabled = get_exec().moe.disable_shared_experts_fusion if not disabled: diff --git a/python/sglang/srt/layers/quantization/compressed_tensors/compressed_tensors.py b/python/sglang/srt/layers/quantization/compressed_tensors/compressed_tensors.py index 6cbae1622..b6a23177e 100644 --- a/python/sglang/srt/layers/quantization/compressed_tensors/compressed_tensors.py +++ b/python/sglang/srt/layers/quantization/compressed_tensors/compressed_tensors.py @@ -69,7 +69,8 @@ from sglang.srt.layers.quantization.unquant import ( UnquantizedFusedMoEMethod, UnquantizedLinearMethod, ) -from sglang.srt.utils import is_cuda, is_hip, is_npu, is_sm100_supported, is_xpu +from sglang.srt.runtime_context import get_platform +from sglang.srt.utils import is_cuda, is_hip, is_npu, is_xpu _is_cuda = is_cuda() _is_npu = is_npu() @@ -838,7 +839,7 @@ class CompressedTensorsConfig(QuantizationConfig): triton_supported = self._is_wna16_triton_moe_supported(weight_quant) use_blackwell_triton = ( moe_backend.is_auto() - and is_sm100_supported() + and get_platform().is_sm100 and triton_supported ) if moe_backend.is_triton() and not triton_supported: diff --git a/python/sglang/srt/layers/quantization/compressed_tensors/schemes/compressed_tensors_w4a4_nvfp4_moe.py b/python/sglang/srt/layers/quantization/compressed_tensors/schemes/compressed_tensors_w4a4_nvfp4_moe.py index 07c11416e..245bc3f19 100644 --- a/python/sglang/srt/layers/quantization/compressed_tensors/schemes/compressed_tensors_w4a4_nvfp4_moe.py +++ b/python/sglang/srt/layers/quantization/compressed_tensors/schemes/compressed_tensors_w4a4_nvfp4_moe.py @@ -10,13 +10,13 @@ from sglang.srt.layers.moe.utils import get_moe_runner_backend from sglang.srt.layers.quantization.compressed_tensors.schemes import ( CompressedTensorsMoEScheme, ) -from sglang.srt.layers.quantization.fp8_utils import is_blackwell_supported from sglang.srt.layers.quantization.utils import ( prepare_static_weights_for_trtllm_fp4_moe, reorder_w1w3_to_w3w1, replace_parameter, swizzle_blockscale, ) +from sglang.srt.runtime_context import get_platform from sglang.srt.utils import set_weight_attrs logger = logging.getLogger(__name__) @@ -33,7 +33,7 @@ if TYPE_CHECKING: class CompressedTensorsW4A4Nvfp4MoE(CompressedTensorsMoEScheme): def __init__(self): - if not is_blackwell_supported(): + if not get_platform().is_blackwell: raise ValueError( "Current platform does not support NVFP4" " quantization. Please use Blackwell and" diff --git a/python/sglang/srt/layers/quantization/fp4_kv_cache_quant_method.py b/python/sglang/srt/layers/quantization/fp4_kv_cache_quant_method.py index d859396c6..611ddccce 100644 --- a/python/sglang/srt/layers/quantization/fp4_kv_cache_quant_method.py +++ b/python/sglang/srt/layers/quantization/fp4_kv_cache_quant_method.py @@ -42,7 +42,7 @@ import torch from torch import Tensor from sglang.srt.layers.quantization.kvfp4_tensor import E2M1_MAX -from sglang.srt.utils.common import is_sm100_supported +from sglang.srt.runtime_context import get_platform class KVCacheAttentionPhase(str, Enum): @@ -416,7 +416,7 @@ class NVFP4KVCacheMethod(KVCacheQuantMethodBase): # The FP4 data type itself is identical on both architectures. # Reference: TRT-LLM FP8QDQLinearMethod.process_weights_after_loading_fused_qkv_linear # https://github.com/NVIDIA/TensorRT-LLM/blob/main/tensorrt_llm/_torch/modules/linear.py - if is_sm100_supported(): + if get_platform().is_sm100: k_scale *= E2M1_MAX v_scale *= E2M1_MAX k_scales_cpu[layer_id] = k_scale diff --git a/python/sglang/srt/layers/quantization/fp4_utils.py b/python/sglang/srt/layers/quantization/fp4_utils.py index 52588dbbf..4f6ef8265 100644 --- a/python/sglang/srt/layers/quantization/fp4_utils.py +++ b/python/sglang/srt/layers/quantization/fp4_utils.py @@ -6,11 +6,13 @@ from typing import Optional import torch -from sglang.srt.runtime_context import get_exec +from sglang.srt.runtime_context import ( + get_exec, + get_platform, +) from sglang.srt.utils.common import ( get_device_capability, is_cuda, - is_sm100_supported, ) from sglang.srt.utils.custom_op import register_custom_op_from_extern @@ -21,7 +23,7 @@ fp4_quantize = None try: from flashinfer import fp4_quantize as _flashinfer_fp4_quantize - _flashinfer_fp4_quantize_backend = "cute-dsl" if is_sm100_supported() else "cuda" + _flashinfer_fp4_quantize_backend = "cute-dsl" if get_platform().is_sm100 else "cuda" def _round_up(x: int, y: int) -> int: return ((x + y - 1) // y) * y @@ -146,7 +148,7 @@ def initialize_fp4_gemm_config() -> None: backend = get_exec().kernel.fp4_gemm_runner_backend if backend == "auto": - if is_sm100_supported(): + if get_platform().is_sm100: backend = "flashinfer_cutedsl" elif is_cuda() and (10, 0) > get_device_capability() >= (8, 0): backend = "marlin" diff --git a/python/sglang/srt/layers/quantization/fp8.py b/python/sglang/srt/layers/quantization/fp8.py index b449988af..73e4744be 100644 --- a/python/sglang/srt/layers/quantization/fp8.py +++ b/python/sglang/srt/layers/quantization/fp8.py @@ -80,11 +80,13 @@ from sglang.srt.layers.quantization.utils import ( requantize_with_max_scale, ) from sglang.srt.layers.utils import copy_or_rebind_param -from sglang.srt.runtime_context import get_parallel +from sglang.srt.runtime_context import ( + get_parallel, + get_platform, +) from sglang.srt.utils import ( cpu_has_amx_support, get_bool_env_var, - is_blackwell_supported, is_cpu, is_cuda, is_flashinfer_available, @@ -92,9 +94,6 @@ from sglang.srt.utils import ( is_hip, is_musa, is_npu, - is_sm90_supported, - is_sm100_supported, - is_sm120_supported, is_xpu, log_info_on_rank0, mxfp8_block_convert_required, @@ -402,7 +401,7 @@ class Fp8Config(QuantizationConfig): if self.is_fp4_experts and get_moe_runner_backend().is_flashinfer_mxfp4(): # SM100 uses TRT-LLM; SM90 uses W4A16 and SM120 uses MXFP8xMXFP4. - if is_sm90_supported() or is_sm120_supported(): + if get_platform().is_sm90 or get_platform().is_sm120: from sglang.srt.layers.quantization.mxfp4_flashinfer_cutlass_moe import ( Mxfp4FlashinferCutlassMoEMethod, ) @@ -793,7 +792,7 @@ class Fp8LinearMethod(LinearMethodBase): scale_u8 = layer.weight_scale_inv.data layer.weight_scale_inv_swizzled = None if n % 64 != 0 or k % 128 != 0: - if not (is_blackwell_supported() and is_flashinfer_available()): + if not (get_platform().is_blackwell and is_flashinfer_available()): raise RuntimeError( f"--fp8-gemm-backend=deep_gemm cannot serve MXFP8 weight shape " f"({n}, {k}) (needs N % 64 == 0 and K % 128 == 0), and this " @@ -1094,7 +1093,9 @@ class Fp8MoEMethod(FusedMoEMethodBase): ), "cutlass_fp8 MoE requires CUDA 12.0+ with SM90 or CUDA 12.4+ with SM89" assert self.block_quant, "cutlass_fp8 MoE requires block quantization" assert ( - is_sm100_supported() or is_sm90_supported() or is_sm120_supported() + get_platform().is_sm100 + or get_platform().is_sm90 + or get_platform().is_sm120 ), "cutlass_fp8 MoE requires SM90, SM100, or SM120 GPUs" @staticmethod @@ -1700,7 +1701,7 @@ class Fp8MoEMethod(FusedMoEMethodBase): layer.w13_weight_scale_inv.format_ue8m0 = True layer.w2_weight_scale_inv.format_ue8m0 = True - if get_moe_a2a_backend().is_megamoe() and is_sm90_supported(): + if get_moe_a2a_backend().is_megamoe() and get_platform().is_sm90: from sglang.srt.layers.moe.mega_moe_sm90 import ( build_sm90_mega_moe_experts_weights, ) @@ -1753,7 +1754,7 @@ class Fp8MoEMethod(FusedMoEMethodBase): def _process_mxfp8_moe_weights(self, layer: Module, quantize: bool = True) -> None: if not ( - (_is_cuda and is_sm100_supported()) or (_is_hip and _is_gfx95_supported) + (_is_cuda and get_platform().is_sm100) or (_is_hip and _is_gfx95_supported) ): raise RuntimeError( "MXFP8 MoE quantization requires SM100 or ROCm gfx95 " diff --git a/python/sglang/srt/layers/quantization/fp8_utils.py b/python/sglang/srt/layers/quantization/fp8_utils.py index 74453d3e5..084dafffe 100755 --- a/python/sglang/srt/layers/quantization/fp8_utils.py +++ b/python/sglang/srt/layers/quantization/fp8_utils.py @@ -26,7 +26,11 @@ from sglang.kernels.ops.quantization.fp8_kernel import ( from sglang.srt.environ import envs from sglang.srt.layers import deep_gemm_wrapper from sglang.srt.layers.quantization.mxfp4_tensor import MXFP4QuantizeUtil -from sglang.srt.runtime_context import get_exec, get_parallel +from sglang.srt.runtime_context import ( + get_exec, + get_parallel, + get_platform, +) from sglang.srt.utils import ( ceil_align, ceil_div, @@ -35,15 +39,11 @@ from sglang.srt.utils import ( get_device_capability, get_device_sm, get_hip_version, - is_blackwell_supported, is_cuda, is_flashinfer_available, is_gfx95_supported, is_hip, is_musa, - is_sm90_supported, - is_sm100_supported, - is_sm120_supported, is_xpu, offloader, ) @@ -56,9 +56,6 @@ _is_hip = is_hip() _is_cuda = is_cuda() _is_xpu = is_xpu() _is_fp8_fnuz = is_fp8_fnuz() -_is_sm90_supported = is_sm90_supported() -_is_sm100_supported = is_sm100_supported() -_is_sm120_supported = is_sm120_supported() _is_gfx95_supported = is_gfx95_supported() _is_musa = is_musa() @@ -381,7 +378,7 @@ FP8_GEMM_RUNNER_BACKEND: Fp8GemmRunnerBackend | None = None @lru_cache(maxsize=1) def flashinfer_per_tensor_fp8_supported() -> bool: return is_flashinfer_available() and ( - is_sm90_supported() or is_sm100_supported() or is_sm120_supported() + get_platform().is_sm90 or get_platform().is_sm100 or get_platform().is_sm120 ) @@ -433,7 +430,7 @@ def _fake_flashinfer_mxfp8_quantize( return q_input, scale -if is_blackwell_supported() and is_flashinfer_available(): +if get_platform().is_blackwell and is_flashinfer_available(): from flashinfer import SfLayout from flashinfer import mm_mxfp8 as _raw_flashinfer_mm_mxfp8 from flashinfer import mxfp8_quantize as _raw_flashinfer_mxfp8_quantize @@ -544,7 +541,7 @@ if is_blackwell_supported() and is_flashinfer_available(): ) -if is_sm90_supported() and is_flashinfer_available(): +if get_platform().is_sm90 and is_flashinfer_available(): # FlashInfer SM90 DeepGEMM with automatic swapAB optimization for small M from flashinfer.gemm import fp8_blockscale_gemm_sm90 @@ -573,7 +570,7 @@ def resolve_mxfp8_dense_gemm_backend() -> Mxfp8DenseGemmBackend: backend = get_fp8_gemm_runner_backend() if backend.is_flashinfer_trtllm(): - if not (_is_sm100_supported and is_flashinfer_available()): + if not (get_platform().is_sm100 and is_flashinfer_available()): raise RuntimeError( "MXFP8 dense GEMM requested via --fp8-gemm-backend=flashinfer_trtllm, " "but that kernel requires SM100/SM103 GPUs and FlashInfer." @@ -582,7 +579,7 @@ def resolve_mxfp8_dense_gemm_backend() -> Mxfp8DenseGemmBackend: if backend.is_flashinfer_cutedsl(): if not ( - is_blackwell_supported() + get_platform().is_blackwell and is_flashinfer_available() and _raw_flashinfer_mm_mxfp8.is_backend_supported( "cute-dsl", get_device_sm() @@ -595,7 +592,7 @@ def resolve_mxfp8_dense_gemm_backend() -> Mxfp8DenseGemmBackend: return Mxfp8DenseGemmBackend.FLASHINFER_CUTEDSL if backend.is_flashinfer_cutlass(): - if not (is_blackwell_supported() and is_flashinfer_available()): + if not (get_platform().is_blackwell and is_flashinfer_available()): raise RuntimeError( "MXFP8 dense GEMM requested via --fp8-gemm-backend=flashinfer_cutlass, " "but that kernel requires Blackwell GPUs and FlashInfer." @@ -614,7 +611,7 @@ def resolve_mxfp8_dense_gemm_backend() -> Mxfp8DenseGemmBackend: if _is_hip and _is_gfx95_supported: return Mxfp8DenseGemmBackend.GFX95_DOT_SCALED - if is_blackwell_supported() and is_flashinfer_available(): + if get_platform().is_blackwell and is_flashinfer_available(): if _raw_flashinfer_mm_mxfp8.is_backend_supported("cute-dsl", get_device_sm()): return Mxfp8DenseGemmBackend.FLASHINFER_CUTEDSL return Mxfp8DenseGemmBackend.FLASHINFER_CUTLASS @@ -712,7 +709,7 @@ def _deepgemm_w8a8_mxfp8_linear_with_fallback( def _dispatch_explicit_backend(backend: Fp8GemmRunnerBackend) -> Callable: """Dispatch based on explicitly selected backend.""" if backend.is_flashinfer_trtllm(): - if not (is_sm100_supported() and is_flashinfer_available()): + if not (get_platform().is_sm100 and is_flashinfer_available()): raise RuntimeError( "FlashInfer FP8 GEMM requested via --fp8-gemm-backend=flashinfer_trtllm, " "but FlashInfer is not available or not supported on this hardware. " @@ -721,7 +718,7 @@ def _dispatch_explicit_backend(backend: Fp8GemmRunnerBackend) -> Callable: return flashinfer_gemm_w8a8_block_fp8_linear_with_fallback elif backend.is_flashinfer_cutlass(): - if not (is_blackwell_supported() and is_flashinfer_available()): + if not (get_platform().is_blackwell and is_flashinfer_available()): raise RuntimeError( "FlashInfer FP8 GEMM requested via --fp8-gemm-backend=flashinfer_cutlass, " "but FlashInfer is not available or not supported on this hardware. " @@ -730,7 +727,7 @@ def _dispatch_explicit_backend(backend: Fp8GemmRunnerBackend) -> Callable: return flashinfer_gemm_w8a8_block_fp8_linear_with_fallback elif backend.is_flashinfer_deepgemm(): - if not (is_sm90_supported() and is_flashinfer_available()): + if not (get_platform().is_sm90 and is_flashinfer_available()): raise RuntimeError( "FlashInfer DeepGEMM with swapAB requested via --fp8-gemm-backend=flashinfer_deepgemm, " "but it's not available. This backend requires Hopper (SM90) GPUs and FlashInfer " @@ -739,7 +736,7 @@ def _dispatch_explicit_backend(backend: Fp8GemmRunnerBackend) -> Callable: return flashinfer_deepgemm_w8a8_block_fp8_linear_with_fallback elif backend.is_cutlass(): - if not is_sm120_supported(): + if not get_platform().is_sm120: raise RuntimeError( "--fp8-gemm-backend=cutlass is deprecated on this hardware. " "Please switch to DeepGEMM or FlashInfer TRTLLM on SM90/SM100." @@ -782,9 +779,9 @@ def _dispatch_auto_backend() -> Callable: if deep_gemm_wrapper.ENABLE_JIT_DEEPGEMM: return deepgemm_w8a8_block_fp8_linear_with_fallback - elif is_blackwell_supported() and is_flashinfer_available(): + elif get_platform().is_blackwell and is_flashinfer_available(): return flashinfer_gemm_w8a8_block_fp8_linear_with_fallback - elif is_sm120_supported(): + elif get_platform().is_sm120: return cutlass_w8a8_block_fp8_linear_with_fallback elif _use_aiter: return aiter_w8a8_block_fp8_linear @@ -797,7 +794,7 @@ def initialize_fp8_gemm_config() -> None: global FP8_GEMM_RUNNER_BACKEND backend = get_exec().kernel.fp8_gemm_runner_backend - if backend == "auto" and is_sm120_supported(): + if backend == "auto" and get_platform().is_sm120: backend = "cutlass" backend = Fp8GemmRunnerBackend(backend) @@ -1839,7 +1836,7 @@ def apply_fp8_linear( use_cutlass_channelwise_gemm and envs.SGLANG_ENABLE_FP8_GEMM_CONFIG_TUNE.get() ) native_scalar_a_scale = use_cutlass_channelwise_gemm and ( - _is_sm90_supported or _is_sm100_supported or _is_sm120_supported + get_platform().is_sm90 or get_platform().is_sm100 or get_platform().is_sm120 ) if input_prequantized: diff --git a/python/sglang/srt/layers/quantization/kvfp4_tensor.py b/python/sglang/srt/layers/quantization/kvfp4_tensor.py index 553cb5f2e..4412b33fa 100644 --- a/python/sglang/srt/layers/quantization/kvfp4_tensor.py +++ b/python/sglang/srt/layers/quantization/kvfp4_tensor.py @@ -17,6 +17,8 @@ from enum import Enum import torch +from sglang.srt.runtime_context import get_platform + class FP4KVCacheRecipe(Enum): MXFP4 = 1 # KVFP4: block-wise scaling @@ -175,14 +177,9 @@ class NVFP4KVQuantizeUtil: block_scales: shape [B, M, N/16], dtype float8_e4m3fn global_scale: passthrough """ - from sglang.srt.utils import ( - is_sm90_supported, - is_sm100_supported, - is_sm120_supported, - ) assert ( - is_sm100_supported() or is_sm120_supported() or is_sm90_supported() + get_platform().is_sm100 or get_platform().is_sm120 or get_platform().is_sm90 ), "NVFP4 KV cache quantize requires SM100/SM120 or SM90 fallback GPU" b, m, n = tensor.shape @@ -202,7 +199,7 @@ class NVFP4KVQuantizeUtil: "NVFP4 global scale tensor must already be on the KV tensor device." ) - if is_sm100_supported() or is_sm120_supported(): + if get_platform().is_sm100 or get_platform().is_sm120: from flashinfer import nvfp4_kv_quantize # nvfp4_kv_quantize takes global_scale directly (not inverted) @@ -249,11 +246,6 @@ class NVFP4KVQuantizeUtil: Returns: Dequantized tensor of shape [B, M, N] """ - from sglang.srt.utils import ( - is_sm90_supported, - is_sm100_supported, - is_sm120_supported, - ) b, m, n_half = quant_tensor.shape @@ -271,7 +263,7 @@ class NVFP4KVQuantizeUtil: "NVFP4 global scale tensor must already be on the KV tensor device." ) - if is_sm100_supported() or is_sm120_supported(): + if get_platform().is_sm100 or get_platform().is_sm120: from flashinfer import nvfp4_kv_dequantize quant_2d = quant_tensor.view(torch.uint8).reshape(b * m, n_half) @@ -282,7 +274,7 @@ class NVFP4KVQuantizeUtil: return output_2d.reshape(b, m, -1) else: assert ( - is_sm90_supported() + get_platform().is_sm90 ), "NVFP4 KV cache dequantize requires SM100/SM120 or SM90 fallback GPU" # Pure PyTorch fallback for SM90 n = n_half * 2 diff --git a/python/sglang/srt/layers/quantization/modelopt_quant.py b/python/sglang/srt/layers/quantization/modelopt_quant.py index 38508638d..f8120da41 100755 --- a/python/sglang/srt/layers/quantization/modelopt_quant.py +++ b/python/sglang/srt/layers/quantization/modelopt_quant.py @@ -42,7 +42,6 @@ from sglang.srt.layers.quantization.fp8_utils import ( can_auto_enable_marlin_fp8, cutlass_fp8_supported, flashinfer_per_tensor_fp8_supported, - is_blackwell_supported, ) from sglang.srt.layers.quantization.kv_cache import BaseKVCacheMethod from sglang.srt.layers.quantization.marlin_utils_fp4 import ( @@ -63,10 +62,10 @@ from sglang.srt.layers.quantization.utils import ( ) from sglang.srt.layers.radix_attention import RadixAttention from sglang.srt.layers.utils import alias_or_bind_derived_param, copy_or_rebind_param +from sglang.srt.runtime_context import get_platform from sglang.srt.utils.common import ( get_device_capability, is_cuda, - is_sm120_supported, round_up, set_weight_attrs, ) @@ -151,7 +150,7 @@ def fp4_gemm( ) -if is_cuda() and (not is_sm120_supported()) and (fp4_quantize is not None): +if is_cuda() and (not get_platform().is_sm120) and (fp4_quantize is not None): @register_fake_if_exists("sgl_kernel::scaled_fp4_quant") def _sgl_kernel_scaled_fp4_quant_fake( @@ -1796,7 +1795,7 @@ class ModelOptFp4LinearMethod(LinearMethodBase): layer.weights_padding_cols = 0 return - if not is_blackwell_supported(): + if not get_platform().is_blackwell: raise ValueError( "ModelOpt NVFP4 native dense GEMM backends require SM100+. " "Use --fp4-gemm-backend marlin on SM80-SM90." @@ -2196,7 +2195,7 @@ class ModelOptNvFp4FusedMoEMethod(FusedMoEMethodBase): use_marlin_fallback = (8, 0) <= capability < (10, 0) else: use_marlin_fallback = moe_runner_backend.is_marlin() - if not is_blackwell_supported() and not use_marlin_fallback: + if not get_platform().is_blackwell and not use_marlin_fallback: raise ValueError( "Current platform does not support NVFP4" " quantization with the selected MoE backend. Please use " diff --git a/python/sglang/srt/layers/quantization/mxfp4.py b/python/sglang/srt/layers/quantization/mxfp4.py index 15e3f2007..369be8973 100644 --- a/python/sglang/srt/layers/quantization/mxfp4.py +++ b/python/sglang/srt/layers/quantization/mxfp4.py @@ -49,7 +49,10 @@ from sglang.srt.layers.quantization.base_config import ( QuantizeMethodBase, ) from sglang.srt.layers.quantization.utils import is_layer_skipped -from sglang.srt.runtime_context import get_exec +from sglang.srt.runtime_context import ( + get_exec, + get_platform, +) from sglang.srt.utils import ( cpu_has_amx_support, get_device_capability, @@ -57,9 +60,6 @@ from sglang.srt.utils import ( is_flashinfer_available, is_gfx95_supported, is_hip, - is_sm90_supported, - is_sm100_supported, - is_sm120_supported, is_triton_kernels_available, next_power_of_2, round_up, @@ -236,13 +236,13 @@ def _swizzle_mxfp4(quant_tensor, scale, num_warps): mx_axis=-2, num_warps=num_warps ) scale_layout_opts = {} - if is_sm100_supported(): + if get_platform().is_sm100: constraints = { "is_persistent": True, "epilogue_subtile": 1, } opt_flags.update_opt_flags_constraints(constraints) - elif is_sm90_supported(): + elif get_platform().is_sm90: constraints = { "split_k": 1, } @@ -413,11 +413,11 @@ class Mxfp4MoEMethod(FusedMoEMethodBase): # (FlashInfer PR #3084, post-0.6.10) self._fi_kernel: Optional[str] = None if self.use_flashinfer: - if is_sm100_supported(): + if get_platform().is_sm100: self._fi_kernel = "trtllm_sm100" - elif is_sm120_supported(): + elif get_platform().is_sm120: self._fi_kernel = "cutlass_sm120" - elif is_sm90_supported(): + elif get_platform().is_sm90: if not _FI_HAS_SM90_CUTLASS_MXFP4: raise RuntimeError( "moe_runner_backend=flashinfer_mxfp4 on SM90 requires the " @@ -466,7 +466,7 @@ class Mxfp4MoEMethod(FusedMoEMethodBase): # DeepGEMM fp8_fp4 grouped GEMM consumes the checkpoint layout # directly (packed e2m1 K-major + ue8m0 g32 scales); no padding. pass - elif is_sm100_supported(): + elif get_platform().is_sm100: if self.use_flashinfer: # FlashInfer trtllm-gen FP4 kernel actual alignment: # intermediate: scale shuffle needs M%128==0 → intermediate%64==0 @@ -614,9 +614,9 @@ class Mxfp4MoEMethod(FusedMoEMethodBase): ) if ( - not is_sm90_supported() - and not is_sm100_supported() - and not is_sm120_supported() + not get_platform().is_sm90 + and not get_platform().is_sm100 + and not get_platform().is_sm120 ): raise RuntimeError("MXFP4 Marlin requires SM90+.") if not check_moe_marlin_supports_layer(layer, 32, allow_tile_padding=True): diff --git a/python/sglang/srt/layers/quantization/mxfp4_flashinfer_cutlass_moe.py b/python/sglang/srt/layers/quantization/mxfp4_flashinfer_cutlass_moe.py index d2e4efab2..8fbdc4aed 100644 --- a/python/sglang/srt/layers/quantization/mxfp4_flashinfer_cutlass_moe.py +++ b/python/sglang/srt/layers/quantization/mxfp4_flashinfer_cutlass_moe.py @@ -14,8 +14,8 @@ import torch from torch.nn import Module from torch.nn.parameter import Parameter +from sglang.srt.runtime_context import get_platform from sglang.srt.utils import is_flashinfer_available, log_info_on_rank0 -from sglang.srt.utils.common import is_sm120_supported # Suppress TRT-LLM CUTLASS trace logs without overriding user configuration. os.environ.setdefault("TLLM_LOG_LEVEL", "INFO") @@ -37,7 +37,7 @@ class Mxfp4FlashinferCutlassMoEMethod: def __init__(self, fp8_method, prefix: str): if not is_flashinfer_available(): raise RuntimeError("Mxfp4FlashinferCutlassMoEMethod requires FlashInfer.") - self._use_mxfp8_act_scaling = is_sm120_supported() + self._use_mxfp8_act_scaling = get_platform().is_sm120 self._fp8 = fp8_method self.prefix = prefix self._swiglu_limit_tensor: torch.Tensor | None = None diff --git a/python/sglang/srt/layers/quantization/mxfp4_flashinfer_trtllm_moe.py b/python/sglang/srt/layers/quantization/mxfp4_flashinfer_trtllm_moe.py index b26a961f5..3d0add5e3 100644 --- a/python/sglang/srt/layers/quantization/mxfp4_flashinfer_trtllm_moe.py +++ b/python/sglang/srt/layers/quantization/mxfp4_flashinfer_trtllm_moe.py @@ -14,15 +14,18 @@ from sglang.srt.distributed.device_communicators.pynccl_allocator import ( ) from sglang.srt.layers.dp_attention import is_allocation_symmetric from sglang.srt.layers.moe.utils import RoutingMethodType -from sglang.srt.runtime_context import get_exec +from sglang.srt.runtime_context import ( + get_exec, + get_platform, +) from sglang.srt.utils import ( is_flashinfer_available, log_info_on_rank0, set_weight_attrs, ) -from sglang.srt.utils.common import is_sm100_supported, next_power_of_2 +from sglang.srt.utils.common import next_power_of_2 -_MXFP8_QUANTIZE_BACKEND = "cute-dsl" if is_sm100_supported() else "cuda" +_MXFP8_QUANTIZE_BACKEND = "cute-dsl" if get_platform().is_sm100 else "cuda" if is_flashinfer_available(): from flashinfer import shuffle_matrix_a, shuffle_matrix_sf_a diff --git a/python/sglang/srt/layers/quantization/mxfp4_marlin_moe.py b/python/sglang/srt/layers/quantization/mxfp4_marlin_moe.py index 631d230c4..585e9d422 100644 --- a/python/sglang/srt/layers/quantization/mxfp4_marlin_moe.py +++ b/python/sglang/srt/layers/quantization/mxfp4_marlin_moe.py @@ -8,8 +8,8 @@ from torch.nn import Module from sglang.srt.layers.moe.moe_runner.marlin import MarlinMoeQuantInfo from sglang.srt.layers.moe.utils import MoeRunnerBackend +from sglang.srt.runtime_context import get_platform from sglang.srt.utils import log_info_on_rank0, round_up, set_weight_attrs -from sglang.srt.utils.common import is_sm90_supported, is_sm120_supported if TYPE_CHECKING: from sglang.srt.layers.moe.token_dispatcher import CombineInput, DispatchOutput @@ -142,7 +142,7 @@ class Mxfp4MarlinMoEMethod: if getattr(layer, "_mega_moe_weights_built", False): return - if not is_sm90_supported() and not is_sm120_supported(): + if not get_platform().is_sm90 and not get_platform().is_sm120: raise RuntimeError("MXFP4 Marlin requires SM90 or SM120.") if not check_moe_marlin_supports_layer(layer, 32, allow_tile_padding=True): diff --git a/python/sglang/srt/layers/quantization/unquant.py b/python/sglang/srt/layers/quantization/unquant.py index 33b7f81b0..2dbd79041 100644 --- a/python/sglang/srt/layers/quantization/unquant.py +++ b/python/sglang/srt/layers/quantization/unquant.py @@ -34,6 +34,7 @@ from sglang.srt.layers.utils import copy_or_rebind_param from sglang.srt.runtime_context import ( get_exec, get_lora, + get_platform, ) from sglang.srt.utils import ( cpu_has_amx_support, @@ -160,10 +161,8 @@ def initialize_bf16_gemm_config(server_args: ServerArgs) -> None: global _flashinfer_pr4266_run_direct_dense global _enable_bf16_splitk_gemm - from sglang.srt.utils import is_sm100_supported - backend_str = server_args.bf16_gemm_backend - if backend_str == "auto" and is_sm100_supported(): + if backend_str == "auto" and get_platform().is_sm100: backend_str = ( "torch" if get_exec().deterministic.enable_deterministic_inference @@ -190,7 +189,7 @@ def initialize_bf16_gemm_config(server_args: ServerArgs) -> None: "--bf16-gemm-backend cutedsl is batch-size dependent and cannot " "be combined with --enable-deterministic-inference" ) - if not is_sm100_supported(): + if not get_platform().is_sm100: raise ValueError( f"--bf16-gemm-backend {backend.value} requires " "SM100/SM103 (Blackwell)" diff --git a/python/sglang/srt/managers/rust_server.py b/python/sglang/srt/managers/rust_server.py index a95862e9d..0b1f20f25 100644 --- a/python/sglang/srt/managers/rust_server.py +++ b/python/sglang/srt/managers/rust_server.py @@ -29,7 +29,9 @@ from sglang.srt.managers.utils import ( msgpack_decode_explained, ) from sglang.srt.runtime_context import ( + get_disagg, get_mm, + get_model, get_observability, get_parallel, get_serving, @@ -179,7 +181,7 @@ class NativeMmHost: self.server_args = server_args self.model_config = model_config # Worker threads == max concurrently-processed mm requests. - self.mm_workers = server_args.mm_processor_worker_num or self.AUTO_MM_WORKERS + self.mm_workers = get_mm().mm_processor_worker_num or self.AUTO_MM_WORKERS # The mapping the Python TokenizerManager builds in # init_tokenizer_and_processor. The caller's already-loaded HF @@ -277,7 +279,7 @@ class NativeMmHost: return ( get_parallel().tp_size > 1 and determine_tensor_transport_mode() != "default" - and not self.server_args.skip_tokenizer_init + and not get_serving().skip_tokenizer_init ) @staticmethod @@ -398,17 +400,17 @@ class RustServer: "ingress has no equivalent). Launch without SGLANG_RUST_SERVER, or " "drop --preferred-sampling-params and send those values per request." ) - http_addr = f"{get_serving().host}:{server_args.port}" + http_addr = f"{get_serving().host}:{get_serving().port}" # Per-DP-rank HTTP port with client load balancing. `None` when DP is off, # so the rank is not conflated with rank 0 of a one-rank group. dp_rank = scheduler.ps.attn_dp_rank if scheduler.ps.dp_size > 1 else None if dp_rank is not None: - http_addr = f"{get_serving().host}:{server_args.port + dp_rank}" + http_addr = f"{get_serving().host}:{get_serving().port + dp_rank}" launch_cores, server_cores = cls._partition_cores( mm_workers=( - (server_args.mm_processor_worker_num or NativeMmHost.AUTO_MM_WORKERS) + (get_mm().mm_processor_worker_num or NativeMmHost.AUTO_MM_WORKERS) if scheduler.model_config.is_multimodal else 0 ) @@ -763,26 +765,26 @@ class RustServer: "null": ext.DisaggregationMode.Null, "prefill": ext.DisaggregationMode.Prefill, "decode": ext.DisaggregationMode.Decode, - }[sa.disaggregation_mode] + }[get_disagg().disaggregation_mode] return ext.ServerArgs( - model_path=sa.model_path, - served_model_name=sa.served_model_name, - tokenizer_path=sa.tokenizer_path, - revision=sa.revision, - load_format=sa.load_format, - weight_version=sa.weight_version, + model_path=get_model().model_path, + served_model_name=get_serving().served_model_name, + tokenizer_path=get_serving().tokenizer_path, + revision=get_model().revision, + load_format=get_model().load_format, + weight_version=get_serving().weight_version, host=get_serving().host, - port=sa.port, + port=get_serving().port, log_level=get_observability().log_level, - log_level_http=sa.log_level_http, - chat_template=sa.chat_template, - tool_call_parser=sa.tool_call_parser, - reasoning_parser=sa.reasoning_parser, - stream_response_default_include_usage=sa.stream_response_default_include_usage, - tokenizer_worker_num=sa.tokenizer_worker_num, - detokenizer_worker_num=sa.detokenizer_worker_num, - skip_tokenizer_init=sa.skip_tokenizer_init, - incremental_streaming_output=sa.incremental_streaming_output, + log_level_http=get_observability().log_level_http, + chat_template=get_serving().chat_template, + tool_call_parser=get_serving().tool_call_parser, + reasoning_parser=get_serving().reasoning_parser, + stream_response_default_include_usage=get_serving().stream_response_default_include_usage, + tokenizer_worker_num=get_serving().tokenizer_worker_num, + detokenizer_worker_num=get_serving().detokenizer_worker_num, + skip_tokenizer_init=get_serving().skip_tokenizer_init, + incremental_streaming_output=get_serving().incremental_streaming_output, disaggregation_mode=disaggregation_mode, model_config=ext.ModelConfig( context_len=mc.context_len, @@ -800,11 +802,11 @@ class RustServer: # `preferred_sampling_params` is deliberately absent: `launch` # refuses to start when it is set, so the Rust server never needs it. preferred_sampling_params=( - json.dumps(sa.preferred_sampling_params) - if sa.preferred_sampling_params is not None + json.dumps(get_serving().preferred_sampling_params) + if get_serving().preferred_sampling_params is not None else None ), - allow_auto_truncate=sa.allow_auto_truncate, + allow_auto_truncate=get_serving().allow_auto_truncate, enable_return_hidden_states=sa.enable_return_hidden_states, # Not a `server_args` field: `TokenizerManager` derives it, and the # rust ingress needs the same number for its total-token check. diff --git a/python/sglang/srt/managers/schedule_batch.py b/python/sglang/srt/managers/schedule_batch.py index eaa943855..7f4733887 100755 --- a/python/sglang/srt/managers/schedule_batch.py +++ b/python/sglang/srt/managers/schedule_batch.py @@ -4,6 +4,7 @@ from sglang.srt.dllm.config import DllmConfig from sglang.srt.model_executor.forward_batch_info import ForwardBatch from sglang.srt.runtime_context import ( get_disagg, + get_parallel, get_schedule, get_serving, get_spec, @@ -137,7 +138,6 @@ from sglang.srt.observability.req_time_stats import ( DPControllerReqTimeStats, SchedulerReqTimeStats, ) -from sglang.srt.runtime_context import get_parallel from sglang.srt.sampling.sampling_batch_info import SamplingBatchInfo from sglang.srt.sampling.sampling_params import SamplingParams from sglang.srt.utils import flatten_nested_list diff --git a/python/sglang/srt/managers/schedule_policy.py b/python/sglang/srt/managers/schedule_policy.py index 460bea0be..a1ffb4f55 100644 --- a/python/sglang/srt/managers/schedule_policy.py +++ b/python/sglang/srt/managers/schedule_policy.py @@ -5,7 +5,10 @@ from array import array from sglang.srt.environ import envs from sglang.srt.managers.prefill_delayer import PrefillDelayerSinglePassExecutor -from sglang.srt.runtime_context import get_disagg +from sglang.srt.runtime_context import ( + get_disagg, + get_schedule, +) from sglang.srt.utils import get_bool_env_var, is_hip _ROUTING_KEY_POLICY_DEBUG_LOG = get_bool_env_var("SGLANG_ROUTING_KEY_POLICY_DEBUG_LOG") @@ -61,7 +64,6 @@ from sglang.srt.mem_cache.multi_ended_allocator import ( UnifiedMambaTokenToKVPoolAllocator, ) from sglang.srt.mem_cache.radix_cache import RadixCache, RadixKey, TreeNode -from sglang.srt.server_args import ServerArgs if TYPE_CHECKING: from sglang.srt.mem_cache.allocator import BaseTokenToKVPoolAllocator @@ -1434,13 +1436,13 @@ class PrefillAdder: return self.budget_state() - def preempt_to_schedule(self, req: Req, server_args: ServerArgs) -> bool: + def preempt_to_schedule(self, req: Req) -> bool: """ Preempt running requests to serve the new request if the priority threshold is met and token count sum is verified. Returns True if preemption was committed, and the new request can be scheduled. """ # Iterate running requests to find preemptible requests - priority_sign = 1 if server_args.schedule_low_priority_values_first else -1 + priority_sign = 1 if get_schedule().schedule_low_priority_values_first else -1 # NOTE: A request finishes in two phases: # 1) update_finish_state + release_kv_cache (in process_batch_result) diff --git a/python/sglang/srt/managers/scheduler.py b/python/sglang/srt/managers/scheduler.py index d66e61077..8dc2f21b9 100644 --- a/python/sglang/srt/managers/scheduler.py +++ b/python/sglang/srt/managers/scheduler.py @@ -29,6 +29,7 @@ from typing import TYPE_CHECKING, Any, Deque, Dict, List, Optional, Set, Tuple, from sglang.srt.runtime_context import ( attention_backends, + get_context, get_device, get_disagg, get_exec, @@ -41,6 +42,7 @@ from sglang.srt.runtime_context import ( get_schedule, get_serving, get_spec, + publish, ) from sglang.srt.utils.common import suppress_noisy_warnings # isort: skip @@ -292,7 +294,6 @@ from sglang.srt.observability.trace import process_tracing_init, trace_set_threa from sglang.srt.parser.reasoning_parser import ReasoningParser from sglang.srt.platforms import current_platform from sglang.srt.plugins import load_plugins -from sglang.srt.runtime_context import get_context, get_spec, publish from sglang.srt.sampling.sampling_batch_info import SamplingBatchInfo from sglang.srt.sampling.sampling_params import TOP_K_ALL from sglang.srt.server_args import PortArgs, ServerArgs, compute_world_size @@ -3526,7 +3527,7 @@ class Scheduler( if running_batch.batch_is_full: if ( not self.enable_priority_preemption - or not adder.preempt_to_schedule(req, self.server_args) + or not adder.preempt_to_schedule(req) ): break diff --git a/python/sglang/srt/managers/scheduler_components/load_publisher.py b/python/sglang/srt/managers/scheduler_components/load_publisher.py index 90a28b73d..b5bada590 100644 --- a/python/sglang/srt/managers/scheduler_components/load_publisher.py +++ b/python/sglang/srt/managers/scheduler_components/load_publisher.py @@ -6,7 +6,7 @@ socket so out-of-process load-aware routers (e.g. sgl-router's router-side in-flight counter. The in-deployment counterpart lives in `sglang.srt.managers.load_snapshot` (SHM / PUSH to node 0), which a router that only knows the worker URL cannot subscribe to; the port is instead -advertised via `/server_info` (`ServerArgs.describe_kv_events_publisher`). +advertised via `/server_info` (`runtime_context.describe_kv_events_publisher`). The payload is a compact tagged subset of `LoadSnapshot` so the wire contract stays fixed as the snapshot grows. diff --git a/python/sglang/srt/managers/scheduler_components/metrics_reporter.py b/python/sglang/srt/managers/scheduler_components/metrics_reporter.py index cffa87d50..092521c40 100644 --- a/python/sglang/srt/managers/scheduler_components/metrics_reporter.py +++ b/python/sglang/srt/managers/scheduler_components/metrics_reporter.py @@ -240,7 +240,7 @@ class SchedulerMetricsReporter: """Initialize Forward Pass Metrics (FPM) publisher if configured.""" self.scheduler.enable_fpm = False if ( - self.scheduler.server_args.enable_forward_pass_metrics + get_observability().enable_forward_pass_metrics and self.scheduler.ps.attn_tp_rank == 0 and self.scheduler.ps.pp_rank == self.scheduler.ps.pp_size - 1 ): @@ -254,7 +254,7 @@ class SchedulerMetricsReporter: else 0 ) self.scheduler._fpm_worker_id = ( - self.scheduler.server_args.forward_pass_metrics_worker_id + get_observability().forward_pass_metrics_worker_id ) base_endpoint = get_observability().forward_pass_metrics_ipc_name if base_endpoint is None: diff --git a/python/sglang/srt/mem_cache/hiradix_cache.py b/python/sglang/srt/mem_cache/hiradix_cache.py index 531595ad1..561599d35 100644 --- a/python/sglang/srt/mem_cache/hiradix_cache.py +++ b/python/sglang/srt/mem_cache/hiradix_cache.py @@ -68,6 +68,7 @@ from sglang.srt.observability.metrics_collector import ( from sglang.srt.runtime_context import ( get_memory, get_observability, + get_parallel, get_serving, ) @@ -104,7 +105,6 @@ class HiRadixCache(RadixCache): # Filled by attach_hybrid_minimax_sparse_pool_to_hiradix_cache. self.token_to_kv_pool_host = None elif isinstance(self.kv_cache, MLATokenToKVPool): - from sglang.srt.runtime_context import get_parallel _parallel = get_parallel() self.token_to_kv_pool_host = MLATokenToKVPoolHost( diff --git a/python/sglang/srt/mem_cache/kv_cache_configurator.py b/python/sglang/srt/mem_cache/kv_cache_configurator.py index bdcd7821c..dfbbd2ada 100644 --- a/python/sglang/srt/mem_cache/kv_cache_configurator.py +++ b/python/sglang/srt/mem_cache/kv_cache_configurator.py @@ -1342,9 +1342,9 @@ class KVCacheConfigurator: PoolCls = HiSparseDSATokenToKVPool from sglang.srt.mem_cache.sparsity import parse_hisparse_config - pool_kwargs["host_to_device_ratio"] = parse_hisparse_config( - self.server_args - ).host_to_device_ratio + pool_kwargs["host_to_device_ratio"] = ( + parse_hisparse_config().host_to_device_ratio + ) elif dsa_cp_layer_shard_rank is not None: # DSA cache layer split: shard KV/indexer layers across CP ranks. from sglang.srt.mem_cache.dsa_cache_layer_split import ( @@ -1759,7 +1759,7 @@ class KVCacheConfigurator: parse_hisparse_config, ) - hisparse_cfg = parse_hisparse_config(self.server_args) + hisparse_cfg = parse_hisparse_config() token_to_kv_pool_allocator = HiSparseTokenToKVPoolAllocator( sizes.max_total_num_tokens, page_size=get_schedule().page_size, diff --git a/python/sglang/srt/mem_cache/mamba_checkpoint_pool.py b/python/sglang/srt/mem_cache/mamba_checkpoint_pool.py index 7a8e71ca4..e1d1237b3 100644 --- a/python/sglang/srt/mem_cache/mamba_checkpoint_pool.py +++ b/python/sglang/srt/mem_cache/mamba_checkpoint_pool.py @@ -53,6 +53,7 @@ from typing import List, Optional import torch from sglang.srt.mem_cache.allocator.mamba import MambaSlotAllocator +from sglang.srt.runtime_context import get_exec logger = logging.getLogger(__name__) @@ -313,7 +314,6 @@ def maybe_init_int8_mamba_checkpoint_pool( allocating, so an oversized ``--int8-mamba-ckpt-size`` fails with an actionable message instead of a cryptic mid-allocation CUDA OOM. """ - from sglang.srt.runtime_context import get_exec try: mamba = get_exec().mamba diff --git a/python/sglang/srt/mem_cache/mamba_radix_cache.py b/python/sglang/srt/mem_cache/mamba_radix_cache.py index 843228045..125a9847b 100644 --- a/python/sglang/srt/mem_cache/mamba_radix_cache.py +++ b/python/sglang/srt/mem_cache/mamba_radix_cache.py @@ -51,6 +51,7 @@ from sglang.srt.mem_cache.multi_ended_allocator import ( from sglang.srt.mem_cache.radix_cache import RadixKey from sglang.srt.mem_cache.utils import split_node_hash_value from sglang.srt.runtime_context import ( + get_parallel, mamba_cache_chunk_size, ) @@ -60,8 +61,6 @@ if TYPE_CHECKING: import logging -from sglang.srt.runtime_context import get_parallel - logger = logging.getLogger(__name__) # Debug-only invariant checks in the Mamba slot-donation path call tensor.item(), diff --git a/python/sglang/srt/mem_cache/pool_host/common.py b/python/sglang/srt/mem_cache/pool_host/common.py index 9068aaabd..ee27e8b30 100644 --- a/python/sglang/srt/mem_cache/pool_host/common.py +++ b/python/sglang/srt/mem_cache/pool_host/common.py @@ -9,6 +9,7 @@ import torch from sglang.srt.environ import envs from sglang.srt.mem_cache.storage.mmap import alloc_mmap +from sglang.srt.runtime_context import get_memory logger = logging.getLogger(__name__) @@ -104,7 +105,6 @@ def get_allocator_from_storage(allocator_type): def get_allocator_type() -> str: """The host-allocator kind the published HiCache configuration asks for.""" - from sglang.srt.runtime_context import get_memory backend = get_memory().hicache_storage_backend if backend == "shm": diff --git a/python/sglang/srt/mem_cache/sparsity/factory.py b/python/sglang/srt/mem_cache/sparsity/factory.py index 7bd141760..132d892ab 100644 --- a/python/sglang/srt/mem_cache/sparsity/factory.py +++ b/python/sglang/srt/mem_cache/sparsity/factory.py @@ -15,6 +15,7 @@ from sglang.srt.mem_cache.sparsity.core.sparse_coordinator import ( SparseConfig, SparseCoordinator, ) +from sglang.srt.runtime_context import get_memory logger = logging.getLogger(__name__) @@ -58,7 +59,7 @@ def _create_backend_adaptor( raise ValueError(f"Unknown attention backend: {backend}") -def _parse_sparse_config(server_args) -> SparseConfig: +def _parse_sparse_config() -> SparseConfig: """Parse hierarchical sparse config from JSON string. Required fields with defaults: top_k (2048), device_buffer_size (2*top_k), @@ -66,7 +67,7 @@ def _parse_sparse_config(server_args) -> SparseConfig: Optional fields (default None): algorithm, backend, min_sparse_prompt_len, page_size. All remaining fields go to sparse_extra_config. """ - extra_config_str = server_args.hisparse_config + extra_config_str = get_memory().hisparse_config if extra_config_str is not None: try: extra_config = json.loads(extra_config_str) @@ -111,9 +112,9 @@ def _parse_sparse_config(server_args) -> SparseConfig: ) -def parse_hisparse_config(server_args) -> SparseConfig: - """Parse hisparse config from server_args, returning defaults if no config provided.""" - return _parse_sparse_config(server_args) +def parse_hisparse_config() -> SparseConfig: + """The hisparse config as resolved, with defaults where none was given.""" + return _parse_sparse_config() def create_sparse_coordinator( @@ -122,10 +123,9 @@ def create_sparse_coordinator( token_to_kv_pool, start_layer: int, end_layer: int, - server_args, **kwargs, ) -> SparseCoordinator: - config = _parse_sparse_config(server_args) + config = _parse_sparse_config() algorithm = _create_sparse_algorithm(config, device, **kwargs) backend_adaptor = _create_backend_adaptor( config.backend, device, algorithm, req_to_token_pool diff --git a/python/sglang/srt/mem_cache/unified_radix_cache.py b/python/sglang/srt/mem_cache/unified_radix_cache.py index 225dffedf..f4371bb41 100644 --- a/python/sglang/srt/mem_cache/unified_radix_cache.py +++ b/python/sglang/srt/mem_cache/unified_radix_cache.py @@ -82,7 +82,11 @@ from sglang.srt.observability.metrics_collector import ( StorageMetrics, StorageMetricsCollector, ) -from sglang.srt.runtime_context import get_memory, get_observability +from sglang.srt.runtime_context import ( + get_memory, + get_model, + get_observability, +) from sglang.srt.session.streaming_session import StreamingSession from sglang.srt.utils.common import ceil_align @@ -432,7 +436,7 @@ class UnifiedRadixCache(BasePrefixCache): ) self.buffer_pipeline = BufferModePipeline( cache=self, - max_context_len=server_args.context_length or 0, + max_context_len=get_model().context_length or 0, swa_window_pages=( swa.full_window_pages if swa is not None and self.tree_core.has_swa_host_pool diff --git a/python/sglang/srt/model_executor/cuda_graph_config.py b/python/sglang/srt/model_executor/cuda_graph_config.py index 8148ef14b..8a7049532 100644 --- a/python/sglang/srt/model_executor/cuda_graph_config.py +++ b/python/sglang/srt/model_executor/cuda_graph_config.py @@ -26,6 +26,8 @@ import json from dataclasses import dataclass, field, replace from typing import Any, Dict, List, Optional +from sglang.srt.runtime_context import get_exec + class Phase: """The two phases of model forward.""" @@ -201,7 +203,6 @@ def check_cuda_graph_backend(phase: str, backend: str) -> bool: """True if cuda_graph_config[phase].backend == backend on the published config. Returns False if the config has not been published yet (e.g. unit tests, early startup).""" - from sglang.srt.runtime_context import get_exec try: cfg = get_exec().graph.cuda_graph_config diff --git a/python/sglang/srt/model_executor/model_runner.py b/python/sglang/srt/model_executor/model_runner.py index 0288f44db..b878d07cf 100644 --- a/python/sglang/srt/model_executor/model_runner.py +++ b/python/sglang/srt/model_executor/model_runner.py @@ -444,7 +444,7 @@ class ModelRunner: # Update deep gemm configure if deep_gemm_wrapper.ENABLE_JIT_DEEPGEMM: - deep_gemm_wrapper.update_deep_gemm_config(gpu_id, server_args) + deep_gemm_wrapper.update_deep_gemm_config(gpu_id) # For hisparse (must be set before initialize() so CUDA graph capture can see it) self.hisparse_coordinator = None @@ -542,7 +542,7 @@ class ModelRunner: self._rearm_eplb_after_elastic_scale() def init_msprobe(self): - self.msprobe_debugger = misc_utils.create_msprobe_debugger(self.server_args) + self.msprobe_debugger = misc_utils.create_msprobe_debugger() def init_weight_updater(self): self.weight_updater = WeightUpdater( @@ -910,7 +910,7 @@ class ModelRunner: ) from sglang.srt.mem_cache.sparsity import parse_hisparse_config - hisparse_cfg = parse_hisparse_config(self.server_args) + hisparse_cfg = parse_hisparse_config() hisparse_top_k = getattr( self.model_config.hf_text_config, "index_topk", hisparse_cfg.top_k ) diff --git a/python/sglang/srt/model_executor/model_runner_components/load_model_utils.py b/python/sglang/srt/model_executor/model_runner_components/load_model_utils.py index d7932a56b..a115b06a0 100644 --- a/python/sglang/srt/model_executor/model_runner_components/load_model_utils.py +++ b/python/sglang/srt/model_executor/model_runner_components/load_model_utils.py @@ -26,6 +26,7 @@ from sglang.srt.model_loader.remote_instance_weight_loader_utils import ( ) from sglang.srt.platforms import current_platform from sglang.srt.runtime_context import ( + get_context, get_exec, get_model, get_observability, @@ -72,7 +73,6 @@ def maybe_downgrade_dtype_for_legacy_gpu(*, model_config: ModelConfig) -> None: logger.info( "Compute capability below sm80. Use float16 due to lack of bfloat16 support." ) - from sglang.srt.runtime_context import get_context # Device-driven, so every runner in the process resolves the same way; # the per-runner truth is model_config.dtype, this is the record. diff --git a/python/sglang/srt/model_executor/model_runner_components/misc_utils.py b/python/sglang/srt/model_executor/model_runner_components/misc_utils.py index 0f4e86cc2..7ce3a09f5 100644 --- a/python/sglang/srt/model_executor/model_runner_components/misc_utils.py +++ b/python/sglang/srt/model_executor/model_runner_components/misc_utils.py @@ -11,13 +11,13 @@ from sglang.srt.configs.model_config import ( from sglang.srt.runtime_context import ( attention_backends, get_context, + get_observability, get_schedule, ) from sglang.srt.server_args import CHUNKED_PREFIX_CACHE_SUPPORTED_ATTENTION_BACKENDS if TYPE_CHECKING: from sglang.srt.configs.model_config import ModelConfig - from sglang.srt.server_args import ServerArgs logger = logging.getLogger(__name__) @@ -48,8 +48,8 @@ def maybe_disable_chunked_prefix_cache( logger.info("Chunked prefix cache is turned on.") -def create_msprobe_debugger(server_args: ServerArgs) -> Optional[Any]: - if server_args.msprobe_dump_config is None: +def create_msprobe_debugger() -> Optional[Any]: + if get_observability().msprobe_dump_config is None: return None try: @@ -62,7 +62,7 @@ def create_msprobe_debugger(server_args: ServerArgs) -> Optional[Any]: return None seed_all(mode=True) - return PrecisionDebugger(config_path=server_args.msprobe_dump_config) + return PrecisionDebugger(config_path=get_observability().msprobe_dump_config) def resolve_pp_proxy_topk_size( diff --git a/python/sglang/srt/model_executor/pool_configurator.py b/python/sglang/srt/model_executor/pool_configurator.py index 534148026..4f6dc72ae 100644 --- a/python/sglang/srt/model_executor/pool_configurator.py +++ b/python/sglang/srt/model_executor/pool_configurator.py @@ -373,7 +373,7 @@ class DefaultPoolConfigurator(MemoryPoolConfigurator): if memory_config.enable_hisparse: from sglang.srt.mem_cache.sparsity import parse_hisparse_config - indexer_ratio = parse_hisparse_config(kvc.server_args).host_to_device_ratio + indexer_ratio = parse_hisparse_config().host_to_device_ratio from sglang.srt.mem_cache.kv_cache_configurator import ( _should_elide_dsa_index_k, @@ -794,9 +794,7 @@ class DSV4PoolConfigurator(MemoryPoolConfigurator): if get_memory().enable_hisparse: from sglang.srt.mem_cache.sparsity import parse_hisparse_config - self.c4_shrink_factor = parse_hisparse_config( - kvc.server_args - ).host_to_device_ratio + self.c4_shrink_factor = parse_hisparse_config().host_to_device_ratio else: self.c4_shrink_factor = 1 assert self.c4_shrink_factor >= 1 diff --git a/python/sglang/srt/model_executor/runner_utils/capture_mode.py b/python/sglang/srt/model_executor/runner_utils/capture_mode.py index 0e394f09a..1c6a28472 100644 --- a/python/sglang/srt/model_executor/runner_utils/capture_mode.py +++ b/python/sglang/srt/model_executor/runner_utils/capture_mode.py @@ -27,6 +27,7 @@ import torch from sglang.srt.model_executor.runner_backend_utils.breakable_cuda_graph.context import ( is_in_breakable_cuda_graph, ) +from sglang.srt.runtime_context import get_flags # Detect whether the current forward pass is in capture mode. is_capture_mode = False @@ -84,7 +85,6 @@ def _set_capture_dsa_variant(variant: Optional[str]) -> None: @contextmanager def model_capture_mode(): global is_capture_mode - from sglang.srt.runtime_context import get_flags # Disable dispose_tensor() during capture: freeing mid-capture records data_ptr()==0 into the graph. is_capture_mode = True diff --git a/python/sglang/srt/model_loader/loader.py b/python/sglang/srt/model_loader/loader.py index dee6486b7..9b784d6d5 100644 --- a/python/sglang/srt/model_loader/loader.py +++ b/python/sglang/srt/model_loader/loader.py @@ -124,7 +124,6 @@ from sglang.srt.model_loader.weight_utils import ( set_runai_streamer_env, ) from sglang.srt.platforms import current_platform -from sglang.srt.runtime_context import get_parallel from sglang.srt.utils import ( get_bool_env_var, get_device_capability, @@ -4388,7 +4387,6 @@ def get_model_loader( if load_config.weight_cache_socket: socket_path = load_config.weight_cache_socket else: - from sglang.srt.runtime_context import get_parallel ps = get_parallel() global_rank = compute_global_rank(ps.tp_size, ps.pp_rank, ps.tp_rank) diff --git a/python/sglang/srt/models/bailing_moe_linear.py b/python/sglang/srt/models/bailing_moe_linear.py index 66de4b208..779d3cc71 100644 --- a/python/sglang/srt/models/bailing_moe_linear.py +++ b/python/sglang/srt/models/bailing_moe_linear.py @@ -58,7 +58,13 @@ from sglang.srt.model_executor.runner import get_is_capture_mode from sglang.srt.model_loader.weight_utils import default_weight_loader from sglang.srt.models.deepseek_v2 import DeepseekV2AttentionMLA, DeepseekV2MLP, _is_hip from sglang.srt.models.utils import WeightsMapper -from sglang.srt.runtime_context import get_device, get_forward, get_parallel, get_stream +from sglang.srt.runtime_context import ( + get_device, + get_forward, + get_parallel, + get_platform, + get_stream, +) from sglang.srt.utils import ( BumpAllocator, add_prefix, @@ -72,7 +78,6 @@ from sglang.srt.utils import ( is_gfx95_supported, is_hip, is_npu, - is_sm100_supported, make_layers, ) from sglang.srt.utils.common import rank0_log @@ -107,7 +112,7 @@ if _is_hip: pass _is_flashinfer_available = is_flashinfer_available() -_is_sm100_supported = is_cuda() and is_sm100_supported() +_is_sm100_supported = is_cuda() and get_platform().is_sm100 class DsV3MLA(DeepseekV2AttentionMLA): diff --git a/python/sglang/srt/models/bailing_moe_v3.py b/python/sglang/srt/models/bailing_moe_v3.py index 14317b338..6b09ba954 100644 --- a/python/sglang/srt/models/bailing_moe_v3.py +++ b/python/sglang/srt/models/bailing_moe_v3.py @@ -83,6 +83,7 @@ from sglang.srt.models.kimi_linear import KimiDeltaAttention from sglang.srt.runtime_context import ( get_forward, get_parallel, + get_platform, get_stream, ) from sglang.srt.utils import ( @@ -91,7 +92,6 @@ from sglang.srt.utils import ( bind_or_assign, is_cuda, is_flashinfer_available, - is_sm100_supported, log_info_on_rank0, make_layers, ) @@ -111,7 +111,7 @@ elif not (_is_cpu and _is_cpu_amx_available): from vllm._custom_ops import awq_dequantize _is_flashinfer_available = is_flashinfer_available() -_is_sm100_supported = is_cuda() and is_sm100_supported() +_is_sm100_supported = is_cuda() and get_platform().is_sm100 class DsV3MLA(DeepseekV2AttentionMLA): diff --git a/python/sglang/srt/models/deepseek_common/amd/deepseek_v4_fused_mhc.py b/python/sglang/srt/models/deepseek_common/amd/deepseek_v4_fused_mhc.py index 81ba655a9..e34f5e46a 100644 --- a/python/sglang/srt/models/deepseek_common/amd/deepseek_v4_fused_mhc.py +++ b/python/sglang/srt/models/deepseek_common/amd/deepseek_v4_fused_mhc.py @@ -5,8 +5,8 @@ import torch import triton from sglang.srt.environ import envs +from sglang.srt.runtime_context import get_platform from sglang.srt.utils import get_bool_env_var, is_gfx95_supported, is_hip -from sglang.srt.utils.common import is_sm120_supported logger = logging.getLogger(__name__) @@ -32,7 +32,7 @@ def _is_fused_mhc_post_pre_enabled() -> bool: return ( envs.SGLANG_OPT_FUSE_MHC_POST_PRE.get() and envs.SGLANG_OPT_USE_TILELANG_MHC_POST.get() - and (envs.SGLANG_OPT_USE_TILELANG_MHC_PRE.get() or is_sm120_supported()) + and (envs.SGLANG_OPT_USE_TILELANG_MHC_PRE.get() or get_platform().is_sm120) ) diff --git a/python/sglang/srt/models/deepseek_common/attention_backend_handler.py b/python/sglang/srt/models/deepseek_common/attention_backend_handler.py index 9a2a60120..b720b0d7f 100644 --- a/python/sglang/srt/models/deepseek_common/attention_backend_handler.py +++ b/python/sglang/srt/models/deepseek_common/attention_backend_handler.py @@ -11,8 +11,11 @@ from sglang.srt.models.deepseek_common.attention_forward_methods.forward_methods AttnForwardMethod, ) from sglang.srt.models.deepseek_common.utils import _is_hip -from sglang.srt.runtime_context import get_exec -from sglang.srt.utils import is_sm100_or_sm110_supported, use_intel_amx_backend +from sglang.srt.runtime_context import ( + get_exec, + get_platform, +) +from sglang.srt.utils import use_intel_amx_backend MHA_ONE_SHOT_SUPPORTED_BACKENDS = ["fa3", "flashinfer", "flashmla"] @@ -157,7 +160,7 @@ def handle_attention_fa4(attn, forward_batch): # flash_attn.cute only implements on SM100/SM110 (not SM120); keep the # pre-existing MHA chunked-KV path elsewhere. Deterministic inference # requires MLA and rejects fa4 on other archs at startup (server_args). - if not is_sm100_or_sm110_supported(): + if not get_platform().is_sm100_or_sm110: return AttnForwardMethod.MHA_CHUNKED_KV if get_exec().deterministic.enable_deterministic_inference: return _dispatch_mla_subtype(attn, forward_batch) diff --git a/python/sglang/srt/models/deepseek_v2.py b/python/sglang/srt/models/deepseek_v2.py index 4202299e5..0e128090e 100644 --- a/python/sglang/srt/models/deepseek_v2.py +++ b/python/sglang/srt/models/deepseek_v2.py @@ -203,6 +203,7 @@ from sglang.srt.runtime_context import ( get_forward, get_model, get_parallel, + get_platform, get_spec, ) from sglang.srt.speculative.spec_info import SpeculativeAlgorithm @@ -764,11 +765,10 @@ class DeepseekV2MoE(nn.Module): from sglang.srt.layers.quantization.modelopt_quant import ( ModelOptFp4LinearMethod, ) - from sglang.srt.utils.common import is_sm100_supported fc1_n = self.shared_experts.gate_up_proj.output_size_per_partition if ( - is_sm100_supported() + get_platform().is_sm100 and isinstance( self.shared_experts.gate_up_proj.quant_method, ModelOptFp4LinearMethod, diff --git a/python/sglang/srt/models/deepseek_v4.py b/python/sglang/srt/models/deepseek_v4.py index 116291a1f..4839ac3eb 100644 --- a/python/sglang/srt/models/deepseek_v4.py +++ b/python/sglang/srt/models/deepseek_v4.py @@ -156,7 +156,13 @@ from sglang.srt.models.deepseek_v2 import ( _is_npu, _is_xpu, ) -from sglang.srt.runtime_context import get_device, get_exec, get_forward, get_parallel +from sglang.srt.runtime_context import ( + get_device, + get_exec, + get_forward, + get_parallel, + get_platform, +) if not _is_hip: from sglang.srt.layers.utils.cp_utils import ( @@ -852,9 +858,7 @@ class MQALayer(MqaAttentionBase): self.alt_streams = None self.alt_streams_indexer = None - from sglang.srt.utils import is_blackwell_supported - - self._multi_stream_bs_limit = 128 if is_blackwell_supported() else 64 + self._multi_stream_bs_limit = 128 if get_platform().is_blackwell else 64 self.compressor = None self.indexer = None diff --git a/python/sglang/srt/models/deepseek_v4_dspark.py b/python/sglang/srt/models/deepseek_v4_dspark.py index 09e182f9e..a90806318 100644 --- a/python/sglang/srt/models/deepseek_v4_dspark.py +++ b/python/sglang/srt/models/deepseek_v4_dspark.py @@ -50,7 +50,10 @@ from sglang.srt.models.dspark import ( project_through_lm_head, run_markov_block, ) -from sglang.srt.runtime_context import get_parallel +from sglang.srt.runtime_context import ( + get_parallel, + get_platform, +) from sglang.srt.speculative.dspark_components.dspark_config import ( get_dspark_sample_from_anchor, parse_dspark_draft_config, @@ -59,7 +62,7 @@ from sglang.srt.speculative.ragged_verify import ( RaggedVerifyMode, read_ragged_verify_mode, ) -from sglang.srt.utils import add_prefix, is_blackwell_supported, is_npu +from sglang.srt.utils import add_prefix, is_npu from sglang.srt.utils.invariants import Bucket, InClosedRange, Invariant, expect logger = logging.getLogger(__name__) @@ -132,7 +135,7 @@ class DSparkAttention(MqaAttentionBase): self._use_fast_kernel = envs.SGLANG_DSPARK_FAST_KERNEL.get() self.alt_streams = alt_streams - self._multi_stream_bs_limit = 128 if is_blackwell_supported() else 64 + self._multi_stream_bs_limit = 128 if get_platform().is_blackwell else 64 if _is_npu: self.register_buffer( "_q_post_norm_weight", diff --git a/python/sglang/srt/models/gemma4_vision.py b/python/sglang/srt/models/gemma4_vision.py index bf8a990dd..d71b0226c 100644 --- a/python/sglang/srt/models/gemma4_vision.py +++ b/python/sglang/srt/models/gemma4_vision.py @@ -29,7 +29,11 @@ from sglang.srt.layers.clippable_linear import ( ) from sglang.srt.layers.layernorm import Gemma4RMSNorm from sglang.srt.layers.quantization.base_config import QuantizationConfig -from sglang.srt.runtime_context import get_mm, get_parallel +from sglang.srt.runtime_context import ( + get_mm, + get_parallel, + get_platform, +) from sglang.srt.utils import ( add_prefix, cpu_has_amx_support, @@ -200,9 +204,8 @@ class Gemma4VisionAttention(nn.Module): if is_cuda(): major, _ = get_device_capability() if major == 9: - from sglang.srt.utils import is_blackwell_supported - if is_blackwell_supported(): + if get_platform().is_blackwell: return "triton_attn" return "fa3" return "triton_attn" diff --git a/python/sglang/srt/models/gpt_oss.py b/python/sglang/srt/models/gpt_oss.py index 1ec1e1962..87f6e4326 100644 --- a/python/sglang/srt/models/gpt_oss.py +++ b/python/sglang/srt/models/gpt_oss.py @@ -72,17 +72,16 @@ from sglang.srt.runtime_context import ( get_exec, get_forward, get_parallel, + get_platform, ) from sglang.srt.utils import ( LazyValue, add_prefix, - is_blackwell_supported, is_cpu, is_cuda, is_flashinfer_available, is_hip, is_npu, - is_sm90_supported, make_layers, ) from sglang.srt.utils.custom_op import register_custom_op @@ -94,7 +93,7 @@ _is_cuda = is_cuda() _is_tinygemm_supported = ( _is_cuda and is_flashinfer_available() - and (is_sm90_supported() or is_blackwell_supported()) + and (get_platform().is_sm90 or get_platform().is_blackwell) ) if _is_tinygemm_supported: diff --git a/python/sglang/srt/models/inkling_common/quantization/config.py b/python/sglang/srt/models/inkling_common/quantization/config.py index 155e4c685..cf16df690 100644 --- a/python/sglang/srt/models/inkling_common/quantization/config.py +++ b/python/sglang/srt/models/inkling_common/quantization/config.py @@ -18,6 +18,7 @@ from sglang.srt.layers.quantization.modelopt_quant import ( ModelOptFp4Config, ModelOptFp4LinearMethod, ) +from sglang.srt.runtime_context import get_parallel from sglang.srt.utils.runai_utils import ObjectStorageModel, is_runai_obj_uri logger = logging.getLogger(__name__) @@ -226,7 +227,6 @@ class InklingModelOptNvfp4Config(ModelOptFp4Config, InklingQuantizationConfigBas def maybe_from_model_config( cls, model_config: ModelConfig ) -> InklingModelOptNvfp4Config | None: - from sglang.srt.runtime_context import get_parallel raw_quant_config = _get_raw_quant_config(model_config) diff --git a/python/sglang/srt/models/kimi_k3.py b/python/sglang/srt/models/kimi_k3.py index 5d9997b11..afe3d9c30 100644 --- a/python/sglang/srt/models/kimi_k3.py +++ b/python/sglang/srt/models/kimi_k3.py @@ -121,8 +121,9 @@ from sglang.srt.multimodal.mm_utils import materialize_multimodal_features from sglang.srt.runtime_context import ( get_exec, get_parallel, + get_platform, ) -from sglang.srt.utils import is_blackwell_supported, is_hip, is_npu, make_layers +from sglang.srt.utils import is_hip, is_npu, make_layers from sglang.srt.utils.common import ( BumpAllocator, add_prefix, @@ -1405,7 +1406,7 @@ class KimiK3DeltaAttention(nn.Module): # Same SM bound rationale as the MLA gate stream. self._bfa_alt_stream = bfa_alt_stream self._bfa_bs_limit = ( - (128 if is_blackwell_supported() else 64) + (128 if get_platform().is_blackwell else 64) if bfa_alt_stream is not None else 0 ) @@ -2050,7 +2051,7 @@ class KimiK3MLAAttention(DeepseekV2AttentionMLA): # on their own and the overlap only adds sync overhead (same # bound as deepseek_v4). self._gate_bs_limit = ( - (128 if is_blackwell_supported() else 64) + (128 if get_platform().is_blackwell else 64) if self._gate_alt_stream is not None else 0 ) diff --git a/python/sglang/srt/models/longcat_flash.py b/python/sglang/srt/models/longcat_flash.py index 4bed1df35..870c94079 100644 --- a/python/sglang/srt/models/longcat_flash.py +++ b/python/sglang/srt/models/longcat_flash.py @@ -88,7 +88,11 @@ from sglang.srt.model_loader.utils import ( ) from sglang.srt.model_loader.weight_utils import default_weight_loader from sglang.srt.models.deepseek_v2 import DeepseekV2AttentionMLA -from sglang.srt.runtime_context import get_parallel, get_stream +from sglang.srt.runtime_context import get_parallel +from sglang.srt.runtime_context import get_parallel as _gp +from sglang.srt.runtime_context import ( + get_stream, +) from sglang.srt.utils import ( BumpAllocator, add_prefix, @@ -141,7 +145,6 @@ def _scmoe_align_rows(t, target): if t is None or t.shape[0] == target: return t from sglang.srt.layers.dp_attention import attn_tp_all_gather_into_tensor as _ag - from sglang.srt.runtime_context import get_parallel as _gp cur = t.shape[0] if target > cur: diff --git a/python/sglang/srt/models/mimo_audio.py b/python/sglang/srt/models/mimo_audio.py index 6e738b3ad..c266e060f 100644 --- a/python/sglang/srt/models/mimo_audio.py +++ b/python/sglang/srt/models/mimo_audio.py @@ -22,7 +22,10 @@ from transformers.models.qwen2.modeling_qwen2 import Qwen2Model from sglang.srt.layers.attention.vision import VisionAttention from sglang.srt.layers.quantization.base_config import QuantizationConfig -from sglang.srt.runtime_context import get_model +from sglang.srt.runtime_context import ( + get_model, + get_parallel, +) logger = logging.getLogger(__name__) @@ -1120,7 +1123,6 @@ class MiMoV2AudioConfig: def _remap_audio_tokenizer_state_dict(state_dict: dict) -> dict: - from sglang.srt.runtime_context import get_parallel tp_size = get_parallel().attn_tp_size tp_rank = get_parallel().attn_tp_rank diff --git a/python/sglang/srt/models/qwen2_moe.py b/python/sglang/srt/models/qwen2_moe.py index 4a93f42f9..ea6cf62f4 100644 --- a/python/sglang/srt/models/qwen2_moe.py +++ b/python/sglang/srt/models/qwen2_moe.py @@ -98,7 +98,12 @@ from sglang.srt.model_executor.runner_backend_utils.breakable_cuda_graph.context is_in_breakable_cuda_graph, ) from sglang.srt.model_loader.weight_utils import default_weight_loader -from sglang.srt.runtime_context import get_exec, get_forward, get_parallel +from sglang.srt.runtime_context import ( + get_exec, + get_forward, + get_parallel, + get_stream, +) from sglang.srt.utils import ( add_prefix, cpu_has_amx_support, @@ -118,7 +123,6 @@ if is_npu(): ) from sglang.srt.environ import envs -from sglang.srt.runtime_context import get_stream from sglang.srt.utils.hf_transformers_utils import get_rope_config _SGLANG_EXPERIMENTAL_LORA_OPTI = envs.SGLANG_EXPERIMENTAL_LORA_OPTI.get() diff --git a/python/sglang/srt/multimodal/processors/base_processor.py b/python/sglang/srt/multimodal/processors/base_processor.py index 00283fede..2cf379ea8 100644 --- a/python/sglang/srt/multimodal/processors/base_processor.py +++ b/python/sglang/srt/multimodal/processors/base_processor.py @@ -40,7 +40,10 @@ from sglang.srt.multimodal.transport.cuda_ipc import ( MmItemMemoryPool, get_mm_feature_pool_size_per_worker, ) -from sglang.srt.runtime_context import get_mm, get_serving +from sglang.srt.runtime_context import ( + get_mm, + get_serving, +) from sglang.srt.utils import ( CLIENT_MEDIA_EXCEPTIONS, configure_media_url_security, @@ -663,6 +666,8 @@ class BaseMultimodalProcessor(ABC): if _is_xpu: return "xpu" if not _is_npu: + # Per-worker placement travels as a constructor argument, and + # this record is that argument. return f"cuda:{server_args.base_gpu_id}" if processor.__class__.__name__ == "MiniMaxVLProcessor": # MiniMax's image/video processors create 10-dim tensors during diff --git a/python/sglang/srt/multimodal/transport/memory_pool.py b/python/sglang/srt/multimodal/transport/memory_pool.py index 3a02435e9..1ba0547c0 100644 --- a/python/sglang/srt/multimodal/transport/memory_pool.py +++ b/python/sglang/srt/multimodal/transport/memory_pool.py @@ -7,6 +7,8 @@ from typing import Optional import torch +from sglang.srt.runtime_context import get_parallel + logger = logging.getLogger(__name__) CONTROL_WORD_BYTES = 4 @@ -57,7 +59,6 @@ def resolve_consumer_rank( return 0 if consumer_rank is None: try: - from sglang.srt.runtime_context import get_parallel # Use the global TP rank. An attention/DCP subgroup rank can alias # another consumer's acknowledgement slot. diff --git a/python/sglang/srt/observability/metrics_collector.py b/python/sglang/srt/observability/metrics_collector.py index a50af6e67..6daebdfec 100644 --- a/python/sglang/srt/observability/metrics_collector.py +++ b/python/sglang/srt/observability/metrics_collector.py @@ -28,6 +28,7 @@ from sglang.srt.model_executor.forward_batch_info import ForwardMode from sglang.srt.observability.utils import exponential_buckets, generate_buckets from sglang.srt.runtime_context import ( exports_expert_balancedness_to_prometheus, + get_context, get_disagg, get_observability, get_schedule, @@ -210,7 +211,6 @@ def resolve_collector_class(role: str, default_cls: type) -> type: An unpublished ``observability`` namespace answers with the default. """ - from sglang.srt.runtime_context import get_context, get_observability if not get_context().is_config_namespace_published("observability"): return default_cls diff --git a/python/sglang/srt/observability/trace.py b/python/sglang/srt/observability/trace.py index c3f1cfbdb..9fad00eaf 100644 --- a/python/sglang/srt/observability/trace.py +++ b/python/sglang/srt/observability/trace.py @@ -25,6 +25,7 @@ from dataclasses import dataclass from typing import Any, Dict, List, Mapping, Optional from sglang.srt.environ import envs +from sglang.srt.runtime_context import get_resources from sglang.srt.utils import get_int_env_var logger = logging.getLogger(__name__) @@ -75,7 +76,6 @@ def extract_trace_headers(headers: Mapping[str, str]) -> Optional[Dict]: def get_global_trace_level() -> int: - from sglang.srt.runtime_context import get_resources resources = get_resources() if resources.trace_level is None: @@ -84,7 +84,6 @@ def get_global_trace_level() -> int: def set_global_trace_level(level: int): - from sglang.srt.runtime_context import get_resources get_resources().trace_level = level diff --git a/python/sglang/srt/runtime_context.py b/python/sglang/srt/runtime_context.py index eaf7efab8..8979e043d 100644 --- a/python/sglang/srt/runtime_context.py +++ b/python/sglang/srt/runtime_context.py @@ -54,7 +54,7 @@ import math import os import sys from contextlib import contextmanager -from typing import TYPE_CHECKING, Any +from typing import TYPE_CHECKING, Any, Dict, Optional if TYPE_CHECKING: from sglang.srt.server_args import ServerArgs @@ -833,6 +833,25 @@ def _build_config_bags(server_args: Any) -> dict: return tops +def _resolved_or_field(server_args: Any, name: str, default: Any) -> Any: + """What resolution decided for `name`, falling back to the field. + + Publishes that carry no config at all (sentinels, mocks) have neither, and + answer with `default`. + """ + if server_args is None: + return default + from sglang.srt.arg_groups.overrides import resolution_result + + decided = resolution_result(server_args, name) + if decided is not None: + return decided + # The default is for the callers that hand over something record-shaped but + # not a record -- the fake configs the context tests publish, and `object()` + # for the sentinel publish. A real ServerArgs always has the field. + return getattr(server_args, name, default) + + class RuntimeContext: """Container for the structured runtime accessors; exposes ``parallel``, ``server_args``, the resolved config namespace bags, ``flags``, @@ -914,9 +933,10 @@ class RuntimeContext: stash, which is what the bags are projected from. """ # Seed the capture tier for the new lifecycle (defaults for sentinel - # and mock publishes, which carry no config). - self.flags.capture.enable_torch_compile = getattr( - server_args, "enable_torch_compile", False + # and mock publishes, which carry no config). Through the resolution, + # not the field: the field is the operator's input. + self.flags.capture.enable_torch_compile = bool( + _resolved_or_field(server_args, "enable_torch_compile", False) ) self._server_args = server_args # The adaptive draft-token bound memoizes on the config *path*, so a new @@ -1145,7 +1165,6 @@ class _ServerArgsOverride: self._prev_parallel_config = ctx.parallel._config self._prev_capture = ctx.flags.capture.enable_torch_compile from sglang.srt.arg_groups.overrides import ( - _apply_fields, declare_late_resolution, ) @@ -1162,18 +1181,20 @@ class _ServerArgsOverride: ) # Declared so the projection sees it; late, because the record is # resolved already and not yet published. - # Underscore names are not fields at all (they seed private property - # caches), so they stay a direct write. - declared = { - name: value for name, value in self._fields.items() if name[0] != "_" - } + # Split on whether the name is a field, not on whether it starts with + # an underscore: `_speculative_draft_quantization_explicitly_set` is a + # real field, and seeding it as a raw attribute would leave the earlier + # declaration authoritative, so `resolution_result` and the bag would + # both keep answering the pre-override value. + fields = set(type(server_args).__dataclass_fields__) + declared = {n: v for n, v in self._fields.items() if n in fields} if declared: declare_late_resolution(server_args, "override_server_args", **declared) - # This hook stands in for a launch: the caller's values are both what - # the operator passed and what resolution decided, so they go on the - # record as well as into the stash. Production late resolution declares - # only -- there the record stays the operator's input. - _apply_fields(server_args, self._fields) + # What is left seeds the record's own private caches (`_model_config` + # and friends), which are not configuration and never were. + seeds = {n: v for n, v in self._fields.items() if n not in fields} + for name, value in seeds.items(): + object.__setattr__(server_args, name, value) ctx.set_server_args(server_args) self._installed = True return server_args @@ -1723,6 +1744,142 @@ def pre_capture_activation_reserve_mb(gpu_mem: float | None) -> float: return reserved_mem +# --- Platform facts ----------------------------------------------------------- +# +# One address for what kind of machine this is, so a reader asks +# `get_platform().is_sm100` and an override is stated once instead of patched +# into every module that imported a probe. True before publish, so the context +# probes when no override is installed; `utils.common` holds the implementation. + +_PLATFORM_PROBES: Dict[str, str] = { + "is_cuda": "is_cuda", + "is_hip": "is_hip", + "is_npu": "is_npu", + "is_xpu": "is_xpu", + "is_musa": "is_musa", + "is_sm90": "is_sm90_supported", + "is_sm100": "is_sm100_supported", + "is_sm100_or_sm110": "is_sm100_or_sm110_supported", + "is_sm120": "is_sm120_supported", + "is_blackwell": "is_blackwell_supported", + "is_hopper_with_cuda_12_3": "is_hopper_with_cuda_12_3", + "has_amx": "cpu_has_amx_support", + "has_flashinfer": "is_flashinfer_available", +} + +# Not yes/no facts, same address. +_PLATFORM_VALUES: Dict[str, str] = { + "device_sm": "get_device_sm", + "device_capability": "get_device_capability", +} + + +class PlatformContext: + """The machine's own facts, with one place to override them. + + Every name maps to a probe in `utils.common`; the probes are + `lru_cache`-d, so reading through here costs a call and a dict lookup + (~26 ns) rather than a device query. + """ + + __slots__ = ("_overrides",) + + def __init__(self) -> None: + object.__setattr__(self, "_overrides", {}) + + def __getattr__(self, name: str) -> Any: + probe = _PLATFORM_PROBES.get(name) or _PLATFORM_VALUES.get(name) + if probe is None: + known = sorted(set(_PLATFORM_PROBES) | set(_PLATFORM_VALUES)) + raise AttributeError( + f"unknown platform fact {name!r}; known: {', '.join(known)}" + ) + overrides = object.__getattribute__(self, "_overrides") + if name in overrides: + return overrides[name] + from sglang.srt.utils import common as _common + + return getattr(_common, probe)() + + def __setattr__(self, name: str, value: Any) -> None: + raise AttributeError( + "platform facts are not assigned; use " + "`sglang.srt.runtime_context.override_platform(...)` so every " + "reader agrees" + ) + + def _install(self, **facts: Any) -> Dict[str, Any]: + unknown = set(facts) - set(_PLATFORM_PROBES) - set(_PLATFORM_VALUES) + if unknown: + raise ValueError(f"unknown platform fact(s): {sorted(unknown)}") + overrides = object.__getattribute__(self, "_overrides") + previous = {k: overrides[k] for k in facts if k in overrides} + missing = [k for k in facts if k not in overrides] + overrides.update(facts) + return {"previous": previous, "missing": missing} + + def _restore(self, saved: Dict[str, Any]) -> None: + overrides = object.__getattribute__(self, "_overrides") + overrides.update(saved["previous"]) + for k in saved["missing"]: + overrides.pop(k, None) + + +_PLATFORM = PlatformContext() + + +def get_platform() -> PlatformContext: + """The machine's facts. Answers before publish, unlike a config bag.""" + return _PLATFORM + + +class _PlatformOverride: + """Scoped platform override: `with override_platform(is_sm100=True): ...`""" + + __slots__ = ("_facts", "_saved") + + def __init__(self, **facts: Any) -> None: + self._facts = facts + self._saved = None + + def install(self) -> PlatformContext: + self._saved = _PLATFORM._install(**self._facts) + return _PLATFORM + + def restore(self) -> None: + if self._saved is not None: + _PLATFORM._restore(self._saved) + self._saved = None + + def __enter__(self) -> PlatformContext: + return self.install() + + def __exit__(self, *exc: Any) -> None: + self.restore() + + def __call__(self, fn: Any) -> Any: + """Also usable as a decorator, like the `patch` it replaces. + + A fresh scope per call: the same object decorating two tests must not + share one saved state. + """ + import functools + + facts = dict(self._facts) + + @functools.wraps(fn) + def wrapper(*args: Any, **kwargs: Any) -> Any: + with _PlatformOverride(**facts): + return fn(*args, **kwargs) + + return wrapper + + +def override_platform(**facts: Any) -> _PlatformOverride: + """Say what kind of machine this is, once, for every reader.""" + return _PlatformOverride(**facts) + + # --- Derived config accessors ------------------------------------------------ # # A few values are computed from several config fields plus the HF config, so @@ -1873,3 +2030,124 @@ def is_ep_joiner() -> bool: def is_ep_scale_joiner() -> bool: """True in a process launched as an elastic-EP scale-up joiner.""" return get_exec().moe.ep_join_mode == "scale" + + +def describe_kv_events_publisher(server_args: Any) -> Optional[dict]: + """Return a structured description of this server's KV-event + publisher, or `None` if publishing is disabled / misconfigured. + + This is the wire contract surfaced under the `kv_events` key on + `/server_info` so KV-aware routers (e.g. the SGLang model + gateway) can subscribe per-worker without operator-supplied port + coordination. The router constructs the per-DP-rank SUB endpoint + as tcp://: for + every rank reported in dp_size. + + Returned descriptor shape: + + { + "publisher": "zmq", + "endpoint_host": "*", # may be a ZMQ wildcard + # ("*", "0.0.0.0", "::"); + # subscribers MUST substitute + # the worker URL's host when + # dialing + "endpoint_port_base": 5557, # base TCP port; per-rank + # port = base + dp_rank + "topic": "", # ZMQ topic prefix on the + # SUB filter (empty = + # subscribe-all) + "block_size": , # subscribers MUST + # hash prompts at this size + "dp_size": , # number of SUB sockets to + # open; not DCP-scaled, as + # DCP shards within a rank + # rather than adding + # publishers + "load_endpoint_port_base": , + # base TCP port of the load + # range (load rank r = base + # + r). Consumers MUST read + # this key, not re-derive + # it; present only when + # --load-publish-endpoint + # opted in and a range + # resolved + "load_topic": "load", # SUB filter for the load + # socket; present iff + # load_endpoint_port_base + # is present + } + + Returns None (i.e. "no publisher to describe") when any of: + + * --kv-events-config is unset / empty / malformed JSON, + * the configured publisher is "null", + * page_size is missing or non-positive (a placeholder + block_size would cause silent KV-cache misses by hashing + prompts at the wrong granularity on the router side), + * the endpoint is not a routable TCP address (inproc:// / + ipc://, missing port, non-integer port, port outside + 1..65535, or a bare unbracketed IPv6 host, which is + ambiguous). + + NOTE for load-socket consumers: pair the load port with the worker's + own URL host, as with the KV SUB endpoints — endpoint_host is a + wildcard ("*", "0.0.0.0", "::") whenever the default packing applies, + so splicing it yields tcp://*:PORT and connects to nothing. + + Reuses parse_advertisable_tcp and resolve_load_pub_range — the same + helpers the scheduler binds through — so the advertisement cannot + drift from the sockets. + """ + from sglang.srt.arg_groups.overrides import resolving_view + + # Lazy import so loading server_args doesn't pull in + # disaggregation / msgspec / zmq at module top level. + from sglang.srt.disaggregation.kv_events import ( + LOAD_TOPIC, + KVEventsConfig, + parse_advertisable_tcp, + resolve_load_pub_range, + ) + + resolved = resolving_view(server_args) + raw = resolved.kv_events_config + page_size = resolved.page_size + if not raw or page_size is None or page_size <= 0: + return None + try: + cfg = KVEventsConfig.from_cli(raw) + except Exception: + # Malformed JSON / schema mismatch. The publisher would + # have failed at server startup; /server_info must + # keep working, so just report "no publisher" to consumers. + return None + if cfg.publisher == "null" or not cfg.endpoint: + return None + resolved_kv = parse_advertisable_tcp(cfg.endpoint) + if resolved_kv is None: + return None + host, port = resolved_kv + + descriptor = { + "publisher": cfg.publisher, + "endpoint_host": host, + "endpoint_port_base": port, + "topic": cfg.topic, + "block_size": resolved.kv_event_block_size, + "dp_size": resolved.dp_size, + } + # Load range, from the same resolver SchedulerLoadPublisher binds + # with (so the two can't drift). The decline reason is logged once at + # startup, not here — this runs per /server_info request. + resolved_range, _reason = resolve_load_pub_range( + kv_endpoint=cfg.endpoint, + replay_endpoint=cfg.replay_endpoint, + dp_size=resolved.dp_size, + load_publish_endpoint=resolved.load_publish_endpoint, + ) + if resolved_range is not None: + descriptor["load_endpoint_port_base"] = resolved_range[1] + descriptor["load_topic"] = LOAD_TOPIC + return descriptor diff --git a/python/sglang/srt/server_args.py b/python/sglang/srt/server_args.py index 7e1102e29..7fa605f24 100644 --- a/python/sglang/srt/server_args.py +++ b/python/sglang/srt/server_args.py @@ -58,6 +58,7 @@ from sglang.srt.arg_groups.overrides import ( remote_instance_transfer_engine_of, resolution_projection, resolving_view, + supports_mamba_cache_extra_buffer, ) from sglang.srt.environ import envs from sglang.srt.function_call.function_call_parser import FunctionCallParser @@ -68,6 +69,11 @@ from sglang.srt.model_executor.cuda_graph_config import ( parse_cuda_graph_config_arg, ) from sglang.srt.parser.reasoning_parser import ReasoningParser +from sglang.srt.runtime_context import ( + get_context, + get_platform, + publish, +) from sglang.srt.speculative.decoupled_spec_io import DecoupledSpecIpcConfig from sglang.srt.utils.common import ( LORA_TARGET_ALL_MODULES, @@ -3776,7 +3782,6 @@ class ServerArgs: return reserved_mem def _support_mamba_cache_extra_buffer(self, model_arch: str): - from sglang.srt.arg_groups.overrides import supports_mamba_cache_extra_buffer return supports_mamba_cache_extra_buffer(self, model_arch) @@ -4163,10 +4168,8 @@ class ServerArgs: # the config bags were projected from. Resolved config changes go to the bags via # get_context().override(source, ...); a value one runner or worker # owns travels as a constructor argument to it. - if ( - getattr(self, "_resolution_finished", False) - and not getattr(self, "_internal_write", False) - and (not name.startswith("_") or name in _underscore_field_names()) + if getattr(self, "_resolution_finished", False) and ( + not name.startswith("_") or name in _underscore_field_names() ): raise AttributeError( f"server_args.{name} assigned after resolution; server_args is " @@ -4223,124 +4226,6 @@ class ServerArgs: cfg = resolving_view(self) return cfg.page_size * self.dcp_size - def describe_kv_events_publisher(self) -> Optional[dict]: - """Return a structured description of this server's KV-event - publisher, or `None` if publishing is disabled / misconfigured. - - This is the wire contract surfaced under the `kv_events` key on - `/server_info` so KV-aware routers (e.g. the SGLang model - gateway) can subscribe per-worker without operator-supplied port - coordination. The router constructs the per-DP-rank SUB endpoint - as tcp://: for - every rank reported in dp_size. - - Returned descriptor shape: - - { - "publisher": "zmq", - "endpoint_host": "*", # may be a ZMQ wildcard - # ("*", "0.0.0.0", "::"); - # subscribers MUST substitute - # the worker URL's host when - # dialing - "endpoint_port_base": 5557, # base TCP port; per-rank - # port = base + dp_rank - "topic": "", # ZMQ topic prefix on the - # SUB filter (empty = - # subscribe-all) - "block_size": , # subscribers MUST - # hash prompts at this size - "dp_size": , # number of SUB sockets to - # open; not DCP-scaled, as - # DCP shards within a rank - # rather than adding - # publishers - "load_endpoint_port_base": , - # base TCP port of the load - # range (load rank r = base - # + r). Consumers MUST read - # this key, not re-derive - # it; present only when - # --load-publish-endpoint - # opted in and a range - # resolved - "load_topic": "load", # SUB filter for the load - # socket; present iff - # load_endpoint_port_base - # is present - } - - Returns None (i.e. "no publisher to describe") when any of: - - * --kv-events-config is unset / empty / malformed JSON, - * the configured publisher is "null", - * page_size is missing or non-positive (a placeholder - block_size would cause silent KV-cache misses by hashing - prompts at the wrong granularity on the router side), - * the endpoint is not a routable TCP address (inproc:// / - ipc://, missing port, non-integer port, port outside - 1..65535, or a bare unbracketed IPv6 host, which is - ambiguous). - - NOTE for load-socket consumers: pair the load port with the worker's - own URL host, as with the KV SUB endpoints — endpoint_host is a - wildcard ("*", "0.0.0.0", "::") whenever the default packing applies, - so splicing it yields tcp://*:PORT and connects to nothing. - - Reuses parse_advertisable_tcp and resolve_load_pub_range — the same - helpers the scheduler binds through — so the advertisement cannot - drift from the sockets. - """ - # Lazy import so loading server_args doesn't pull in - # disaggregation / msgspec / zmq at module top level. - from sglang.srt.disaggregation.kv_events import ( - LOAD_TOPIC, - KVEventsConfig, - parse_advertisable_tcp, - resolve_load_pub_range, - ) - - resolved = resolving_view(self) - raw = resolved.kv_events_config - page_size = resolved.page_size - if not raw or page_size is None or page_size <= 0: - return None - try: - cfg = KVEventsConfig.from_cli(raw) - except Exception: - # Malformed JSON / schema mismatch. The publisher would - # have failed at server startup; /server_info must - # keep working, so just report "no publisher" to consumers. - return None - if cfg.publisher == "null" or not cfg.endpoint: - return None - resolved_kv = parse_advertisable_tcp(cfg.endpoint) - if resolved_kv is None: - return None - host, port = resolved_kv - - descriptor = { - "publisher": cfg.publisher, - "endpoint_host": host, - "endpoint_port_base": port, - "topic": cfg.topic, - "block_size": resolved.kv_event_block_size, - "dp_size": resolved.dp_size, - } - # Load range, from the same resolver SchedulerLoadPublisher binds - # with (so the two can't drift). The decline reason is logged once at - # startup, not here — this runs per /server_info request. - resolved_range, _reason = resolve_load_pub_range( - kv_endpoint=cfg.endpoint, - replay_endpoint=cfg.replay_endpoint, - dp_size=resolved.dp_size, - load_publish_endpoint=self.load_publish_endpoint, - ) - if resolved_range is not None: - descriptor["load_endpoint_port_base"] = resolved_range[1] - descriptor["load_topic"] = LOAD_TOPIC - return descriptor - # -------------------------------------------------------------------------- # Module-level ServerArgs helpers and runtime shims. @@ -4381,12 +4266,11 @@ def m3_fp8_attn_gemm_enabled(args) -> bool: bf16 q) without having to move off trtllm_mha. """ from sglang.srt.environ import envs - from sglang.srt.utils.common import is_sm100_supported return ( args.kv_cache_dtype == "fp8_e4m3" and args.attention_backend == "trtllm_mha" - and is_sm100_supported() + and get_platform().is_sm100 and not envs.SGLANG_DISABLE_M3_FP8_ATTN_GEMM.get() ) @@ -4417,7 +4301,6 @@ def _underscore_field_names() -> frozenset: def set_global_server_args_for_scheduler(server_args: ServerArgs): """Legacy publish shim (role=scheduler) — prefer ``runtime_context.publish(server_args, role=...)`` in new code.""" - from sglang.srt.runtime_context import publish publish(server_args, role="scheduler") @@ -4425,7 +4308,6 @@ def set_global_server_args_for_scheduler(server_args: ServerArgs): def set_global_server_args_for_tokenizer(server_args: ServerArgs): """Legacy publish shim (role=tokenizer). Not aliased to the scheduler shim: the process role differs.""" - from sglang.srt.runtime_context import publish publish(server_args, role="tokenizer") @@ -4433,7 +4315,6 @@ def set_global_server_args_for_tokenizer(server_args: ServerArgs): def get_global_server_args() -> ServerArgs: """Legacy accessor shim — prefer ``get_server_args()`` from ``sglang.srt.runtime_context`` in new code.""" - from sglang.srt.runtime_context import get_context return get_context().server_args diff --git a/python/sglang/srt/speculative/dflash_utils.py b/python/sglang/srt/speculative/dflash_utils.py index db852e486..4489afb8c 100644 --- a/python/sglang/srt/speculative/dflash_utils.py +++ b/python/sglang/srt/speculative/dflash_utils.py @@ -17,6 +17,7 @@ from sglang.srt.layers.sampler import ( top_p_normalize_probs_torch, ) from sglang.srt.managers.schedule_batch import Req +from sglang.srt.runtime_context import get_spec from sglang.srt.speculative.spec_utils import sample_simulated_acc_len from sglang.srt.utils import is_cuda, is_hip, is_musa, is_npu @@ -913,11 +914,9 @@ def compute_dflash_sampling_correct_drafts_and_bonus( ) if threshold_single is None: - from sglang.srt.runtime_context import get_spec threshold_single = get_spec().speculative_accept_threshold_single if threshold_acc is None: - from sglang.srt.runtime_context import get_spec threshold_acc = get_spec().speculative_accept_threshold_acc threshold_single = float(threshold_single) diff --git a/python/sglang/srt/state_capturer/indexer_topk.py b/python/sglang/srt/state_capturer/indexer_topk.py index 0360dacd1..b20b5e450 100644 --- a/python/sglang/srt/state_capturer/indexer_topk.py +++ b/python/sglang/srt/state_capturer/indexer_topk.py @@ -6,7 +6,12 @@ import pybase64 import torch from sglang.srt.configs.model_config import ModelConfig, get_num_indexer_layers -from sglang.srt.runtime_context import get_exec, get_parallel, get_schedule +from sglang.srt.runtime_context import ( + get_exec, + get_parallel, + get_resources, + get_schedule, +) from sglang.srt.state_capturer.base import BaseTopkCapturer logger = logging.getLogger(__name__) @@ -42,13 +47,11 @@ class IndexerTopkCapturer(BaseTopkCapturer): def get_global_indexer_capturer() -> Optional[IndexerTopkCapturer]: - from sglang.srt.runtime_context import get_resources return get_resources().indexer_capturer def set_global_indexer_capturer(capturer: Optional[IndexerTopkCapturer]): - from sglang.srt.runtime_context import get_resources get_resources().indexer_capturer = capturer diff --git a/python/sglang/srt/state_capturer/routed_experts.py b/python/sglang/srt/state_capturer/routed_experts.py index cebaa3a3d..57f512d8b 100644 --- a/python/sglang/srt/state_capturer/routed_experts.py +++ b/python/sglang/srt/state_capturer/routed_experts.py @@ -15,6 +15,7 @@ from sglang.srt.model_executor.forward_batch_info import ForwardBatch from sglang.srt.runtime_context import ( get_exec, get_parallel, + get_resources, get_schedule, ) from sglang.srt.state_capturer.base import BaseTopkCapturer @@ -134,13 +135,11 @@ class RoutedExpertsCapturer(BaseTopkCapturer): def get_global_experts_capturer() -> Optional[RoutedExpertsCapturer]: - from sglang.srt.runtime_context import get_resources return get_resources().experts_capturer def set_global_experts_capturer(capturer: Optional[RoutedExpertsCapturer]): - from sglang.srt.runtime_context import get_resources get_resources().experts_capturer = capturer diff --git a/python/sglang/srt/utils/common.py b/python/sglang/srt/utils/common.py index 3ac8ace65..641fd9fa7 100644 --- a/python/sglang/srt/utils/common.py +++ b/python/sglang/srt/utils/common.py @@ -100,7 +100,10 @@ from sglang.srt.observability.func_timer import enable_func_timer from sglang.srt.platforms import current_platform from sglang.srt.runtime_context import ( get_exec, + get_flags, + get_model, get_parallel, + get_spec, ) from sglang.srt.utils.video_decoder import _BACKEND, VideoDecoderWrapper @@ -2336,6 +2339,8 @@ def configure_logger(server_args, prefix: str = ""): maybe_ms = ".%(msecs)03d" if envs.SGLANG_LOG_MS.get() else "" format = f"[%(asctime)s{maybe_ms}{prefix}] %(message)s" logging.basicConfig( + # Runs before publish, and for multimodal_gen's ServerArgs, which + # never publishes these bags -- so the record, not the bag. level=getattr(logging, server_args.log_level.upper()), format=format, datefmt="%Y-%m-%d %H:%M:%S", @@ -3688,8 +3693,6 @@ def dispose_tensor(x: torch.Tensor): if is_in_tc_piecewise_cuda_graph() or is_in_breakable_cuda_graph(): return - from sglang.srt.runtime_context import get_flags - if get_flags().capture.disable_dispose_tensor: return @@ -3723,7 +3726,6 @@ def require_mlp_tp_gather(): Check if the input of MLP is obtained by all-gather rather than all-reduce. This only happens when each MLP TP group contains multiple attention DP groups. """ from sglang.srt.layers.moe.utils import get_moe_a2a_backend - from sglang.srt.runtime_context import get_exec, get_parallel # elastic-EP scale-up rewrites dp_size on the published config if get_parallel().enable_dp_attention: @@ -3783,7 +3785,6 @@ def require_attn_tp_gather(): # and do not consume the upstream gathered_buffer. Without this, the # cuda graph runner pads num_tokens to attn_tp_size, which can cause # autotuners to pick suboptimal kernel variants at small batches. - from sglang.srt.runtime_context import get_parallel if get_parallel().disable_attn_tp_gather: return False @@ -3807,7 +3808,6 @@ def require_gathered_buffer(): def require_mlp_sync(): - from sglang.srt.runtime_context import get_parallel return get_parallel().enable_dp_attention or require_gathered_buffer() @@ -4643,7 +4643,6 @@ def reserve_rope_cache_for_long_sequences(model, model_config, logger=None): resolution's answers. """ from sglang.srt.environ import envs - from sglang.srt.runtime_context import get_model, get_spec SAFETY_FACTOR = envs.SGLANG_SPEC_EXPANSION_SAFETY_FACTOR.get() MARGIN = envs.SGLANG_ROPE_CACHE_SAFETY_MARGIN.get() diff --git a/python/sglang/srt/utils/cuda_vmm_transport_utils.py b/python/sglang/srt/utils/cuda_vmm_transport_utils.py index 399f145a7..3216a06a1 100644 --- a/python/sglang/srt/utils/cuda_vmm_transport_utils.py +++ b/python/sglang/srt/utils/cuda_vmm_transport_utils.py @@ -18,6 +18,7 @@ from sglang.srt.managers.schedule_batch import ( from sglang.srt.runtime_context import ( get_mm, get_parallel, + get_serving, ) from sglang.srt.utils.cuda_ipc_transport_utils import ( DEFER_CUDA_IPC_FEATURE_RECONSTRUCTION_KEY, @@ -941,14 +942,15 @@ class CudaVmmFeatureTransport: ) per_worker_pool_size = get_mm_feature_pool_size_per_worker( - MM_FEATURE_CACHE_SIZE, server_args.tokenizer_worker_num + MM_FEATURE_CACHE_SIZE, get_serving().tokenizer_worker_num ) self.pool = CudaVmmMemoryPool( memory_size=per_worker_pool_size, recycle_interval=MM_ITEM_MEMORY_POOL_RECYCLE_INTERVAL, + # Per-worker placement; policy above reads the bags. base_gpu_id=server_args.base_gpu_id, consumer_count=get_vmm_feature_consumer_count(), - allow_posix_fallback=server_args.nnodes == 1, + allow_posix_fallback=get_parallel().nnodes == 1, ) def prepare_for_dispatch( diff --git a/python/sglang/srt/weight_cache/ipc_loader.py b/python/sglang/srt/weight_cache/ipc_loader.py index 4b7e90897..b4de2b825 100644 --- a/python/sglang/srt/weight_cache/ipc_loader.py +++ b/python/sglang/srt/weight_cache/ipc_loader.py @@ -22,6 +22,7 @@ from sglang.srt.model_loader.loader import ( BaseModelLoader, _initialize_model, ) +from sglang.srt.runtime_context import get_exec, get_parallel from .protocol import ( CacheConfig, @@ -495,7 +496,6 @@ class IpcModelLoader(BaseModelLoader): try: # Build engine's config fingerprint from sglang.srt.layers.dp_attention import get_moe_cp_size - from sglang.srt.runtime_context import get_exec, get_parallel ps = get_parallel() tp_size = ps.tp_size diff --git a/test/registered/unit/configs/test_multimodal_piecewise_cuda_graph.py b/test/registered/unit/configs/test_multimodal_piecewise_cuda_graph.py index 7f016d5fc..13d03bc6a 100644 --- a/test/registered/unit/configs/test_multimodal_piecewise_cuda_graph.py +++ b/test/registered/unit/configs/test_multimodal_piecewise_cuda_graph.py @@ -25,6 +25,7 @@ from sglang.srt.model_executor.forward_batch_info import ( from sglang.srt.model_executor.runner.prefill_cuda_graph_runner import ( PrefillCudaGraphRunner, ) +from sglang.srt.runtime_context import override_platform from sglang.srt.server_args import ServerArgs from sglang.test.ci.ci_register import register_cpu_ci from sglang.test.test_utils import CustomTestCase @@ -84,16 +85,12 @@ class TestMultimodalPiecewiseCudaGraph(CustomTestCase): ) args._cuda_graph_config_locked = set() - with ( - patch( - "sglang.srt.arg_groups.cuda_graph_hook" - ".disable_tc_piecewise_cudagraph_if_incompatible" - ) as disable_if_incompatible, - patch( - "sglang.srt.arg_groups.overrides.attention_backends_of", - return_value=("fa3", "fa3"), - ), - ): + args.attention_backend = "fa3" + + with patch( + "sglang.srt.arg_groups.cuda_graph_hook" + ".disable_tc_piecewise_cudagraph_if_incompatible" + ) as disable_if_incompatible: apply_cuda_graph_compatibility(args) self.assertEqual( @@ -118,11 +115,9 @@ class TestMultimodalPiecewiseCudaGraph(CustomTestCase): ) args._cuda_graph_config_locked = set() - with patch( - "sglang.srt.arg_groups.overrides.attention_backends_of", - return_value=("trtllm_mla", "trtllm_mla"), - ): - apply_cuda_graph_compatibility(args) + args.attention_backend = "trtllm_mla" + + apply_cuda_graph_compatibility(args) self.assertEqual( resolution_result(args, "cuda_graph_config").prefill.backend, @@ -136,11 +131,9 @@ class TestMultimodalPiecewiseCudaGraph(CustomTestCase): ) args._cuda_graph_config_locked = {(Phase.PREFILL, "backend")} - with patch( - "sglang.srt.arg_groups.overrides.attention_backends_of", - return_value=("trtllm_mla", "trtllm_mla"), - ): - apply_cuda_graph_compatibility(args) + args.attention_backend = "trtllm_mla" + + apply_cuda_graph_compatibility(args) self.assertEqual( resolution_result(args, "cuda_graph_config").prefill.backend, @@ -183,7 +176,7 @@ class TestMultimodalPiecewiseCudaGraph(CustomTestCase): args.disable_radix_cache = False args.chunked_prefill_size = 2048 - with (patch("sglang.srt.arg_groups.model_hook.is_cuda", return_value=True),): + with (override_platform(is_cuda=True),): handle_model_capability_adjustments(args) self.assertTrue(resolution_result(args, "disable_radix_cache")) diff --git a/test/registered/unit/constrained/test_base_grammar_backend.py b/test/registered/unit/constrained/test_base_grammar_backend.py index ee3a37627..94dd29451 100644 --- a/test/registered/unit/constrained/test_base_grammar_backend.py +++ b/test/registered/unit/constrained/test_base_grammar_backend.py @@ -20,6 +20,7 @@ import unittest from concurrent.futures import Future from unittest.mock import MagicMock, patch +from sglang.srt.arg_groups.overrides import resolution_result from sglang.srt.constrained.base_grammar_backend import ( GRAMMAR_BACKEND_REGISTRY, BaseGrammarBackend, @@ -355,7 +356,10 @@ class TestCreateGrammarBackend(unittest.TestCase): self.assertEqual( get_context().resolved_server_args_dict()["grammar_backend"], "none" ) - self.assertEqual(server_args.grammar_backend, "xgrammar") + # The record is not written any more: what the caller asked for is a + # declaration on it, and the runtime fallback to "none" lives in the bag + # (asserted above). The two are meant to differ here. + self.assertEqual(resolution_result(server_args, "grammar_backend"), "xgrammar") @patch("sglang.srt.constrained.llguidance_backend.GuidanceBackend") def test_llguidance_backend(self, mock_guidance_cls): diff --git a/test/registered/unit/layers/attention/test_kda_helion_dispatcher.py b/test/registered/unit/layers/attention/test_kda_helion_dispatcher.py index b1e1186ce..dc4cd87e6 100644 --- a/test/registered/unit/layers/attention/test_kda_helion_dispatcher.py +++ b/test/registered/unit/layers/attention/test_kda_helion_dispatcher.py @@ -8,6 +8,7 @@ from sglang.srt.layers.attention.linear.kda_backend import KDAKernelDispatcher from sglang.srt.layers.attention.linear.kernels.kda_helion import HelionKDAKernel from sglang.srt.layers.attention.linear.kernels.kda_triton import TritonKDAKernel from sglang.srt.layers.attention.linear.utils import LinearAttnKernelBackend +from sglang.srt.runtime_context import override_platform from sglang.srt.server_args import ServerArgs from sglang.test.ci.ci_register import register_cpu_ci @@ -160,11 +161,8 @@ class TestHelionKDADispatcher(unittest.TestCase): def test_replayssm_accepts_helion_and_rejects_other_backends(self): with ( - patch( - "sglang.srt.arg_groups.attention_hook.is_sm100_supported", - return_value=False, - ), - patch("sglang.srt.arg_groups.attention_hook.is_cuda", return_value=False), + override_platform(is_sm100=False), + override_platform(is_cuda=False), ): helion_args = ServerArgs( model_path="dummy", @@ -188,11 +186,8 @@ class TestHelionKDADispatcher(unittest.TestCase): mamba_ssm_dtype="bfloat16", ) with ( - patch( - "sglang.srt.arg_groups.attention_hook.is_sm100_supported", - return_value=True, - ), - patch("sglang.srt.arg_groups.attention_hook.is_cuda", return_value=False), + override_platform(is_sm100=True), + override_platform(is_cuda=False), ): handle_linear_attn_backend(args) diff --git a/test/registered/unit/layers/quantization/test_compressed_tensors_wna16_moe_no_linear.py b/test/registered/unit/layers/quantization/test_compressed_tensors_wna16_moe_no_linear.py index 1721ce2ef..5d712b1e8 100644 --- a/test/registered/unit/layers/quantization/test_compressed_tensors_wna16_moe_no_linear.py +++ b/test/registered/unit/layers/quantization/test_compressed_tensors_wna16_moe_no_linear.py @@ -12,6 +12,7 @@ from sglang.srt.layers.quantization.compressed_tensors.schemes import ( CompressedTensorsWNA16MoE, CompressedTensorsWNA16TritonMoE, ) +from sglang.srt.runtime_context import override_platform from sglang.test.ci.ci_register import register_cpu_ci from sglang.test.test_utils import CustomTestCase @@ -90,9 +91,7 @@ class TestWNA16MoENoLinearGroup(CustomTestCase): "get_moe_runner_backend", return_value=MoeRunnerBackend.AUTO, ), - mock.patch.object( - compressed_tensors, "is_sm100_supported", return_value=True - ), + override_platform(is_sm100=True), ): scheme = quant_config.get_moe_scheme( torch.nn.Module(), layer_name=EXPERTS_LAYER @@ -120,9 +119,7 @@ class TestWNA16MoENoLinearGroup(CustomTestCase): "get_moe_runner_backend", return_value=MoeRunnerBackend.AUTO, ), - mock.patch.object( - compressed_tensors, "is_sm100_supported", return_value=True - ), + override_platform(is_sm100=True), ): scheme = quant_config.get_moe_scheme( torch.nn.Module(), layer_name=EXPERTS_LAYER @@ -157,9 +154,7 @@ class TestWNA16MoENoLinearGroup(CustomTestCase): "get_moe_runner_backend", return_value=MoeRunnerBackend.MARLIN, ), - mock.patch.object( - compressed_tensors, "is_sm100_supported", return_value=True - ), + override_platform(is_sm100=True), ): scheme = quant_config.get_moe_scheme( torch.nn.Module(), layer_name=EXPERTS_LAYER @@ -178,9 +173,7 @@ class TestWNA16MoENoLinearGroup(CustomTestCase): "get_moe_runner_backend", return_value=MoeRunnerBackend.AUTO, ), - mock.patch.object( - compressed_tensors, "is_sm100_supported", return_value=True - ), + override_platform(is_sm100=True), ): scheme = quant_config.get_moe_scheme( torch.nn.Module(), layer_name=EXPERTS_LAYER diff --git a/test/registered/unit/layers/quantization/test_mxfp4_sm120_cutlass.py b/test/registered/unit/layers/quantization/test_mxfp4_sm120_cutlass.py index 435c6d7fe..cd23043a6 100644 --- a/test/registered/unit/layers/quantization/test_mxfp4_sm120_cutlass.py +++ b/test/registered/unit/layers/quantization/test_mxfp4_sm120_cutlass.py @@ -11,6 +11,7 @@ from types import SimpleNamespace import pytest import torch +from sglang.srt.runtime_context import override_platform from sglang.test.ci.ci_register import register_cuda_ci register_cuda_ci(est_time=120, stage="base-b", runner_config="1-gpu-small") @@ -81,10 +82,12 @@ def test_cutlass_adapter_import_does_not_require_flashinfer(monkeypatch): sys.modules[module_name] = cached_module -def test_dsv4_sm120_load_contract(monkeypatch): +def test_dsv4_sm120_load_contract(monkeypatch, request): import sglang.srt.layers.quantization.mxfp4_flashinfer_cutlass_moe as adapter_module - monkeypatch.setattr(adapter_module, "is_sm120_supported", lambda: True) + platform = override_platform(is_sm120=True) + platform.install() + request.addfinalizer(platform.restore) captured = {} diff --git a/test/registered/unit/layers/test_flashinfer_comm_fusion.py b/test/registered/unit/layers/test_flashinfer_comm_fusion.py index 5fdc935b8..63d607507 100644 --- a/test/registered/unit/layers/test_flashinfer_comm_fusion.py +++ b/test/registered/unit/layers/test_flashinfer_comm_fusion.py @@ -5,7 +5,7 @@ from unittest.mock import MagicMock, patch import torch from sglang.srt.layers import flashinfer_comm_fusion as fusion -from sglang.srt.runtime_context import get_parallel +from sglang.srt.runtime_context import get_parallel, override_platform from sglang.test.ci.ci_register import register_cuda_ci from sglang.test.test_utils import CustomTestCase @@ -101,7 +101,7 @@ class TestFlashInferCommFusion(CustomTestCase): multi_node = ("auto", True) # Blackwell: mnnvl on both single-node and multi-node. - with patch.object(fusion, "is_sm100_supported", return_value=True): + with override_platform(is_sm100=True): self.assertEqual( fusion._resolve_backend(*single_node), "mnnvl", @@ -110,8 +110,8 @@ class TestFlashInferCommFusion(CustomTestCase): # SM90: auto uses trtllm on single-node, multi-node is unsupported. with ( - patch.object(fusion, "is_sm100_supported", return_value=False), - patch.object(fusion, "is_sm90_supported", return_value=True), + override_platform(is_sm100=False), + override_platform(is_sm90=True), ): self.assertEqual( fusion._resolve_backend(*single_node), @@ -125,8 +125,8 @@ class TestFlashInferCommFusion(CustomTestCase): for arch in ("pre_sm90", "post_sm10x"): with ( self.subTest(arch=arch), - patch.object(fusion, "is_sm100_supported", return_value=False), - patch.object(fusion, "is_sm90_supported", return_value=False), + override_platform(is_sm100=False), + override_platform(is_sm90=False), ): with self.assertRaises(ValueError): fusion._resolve_backend(*single_node) @@ -140,8 +140,8 @@ class TestFlashInferCommFusion(CustomTestCase): multi_node_trtllm = ("trtllm", True) with ( - patch.object(fusion, "is_sm100_supported", return_value=False), - patch.object(fusion, "is_sm90_supported", return_value=True), + override_platform(is_sm100=False), + override_platform(is_sm90=True), ): self.assertEqual( fusion._resolve_backend(*single_node_mnnvl), @@ -156,7 +156,7 @@ class TestFlashInferCommFusion(CustomTestCase): with self.assertRaises(ValueError): fusion._resolve_backend(*multi_node_trtllm) - with patch.object(fusion, "is_sm100_supported", return_value=True): + with override_platform(is_sm100=True): self.assertEqual( fusion._resolve_backend(*multi_node_mnnvl), "mnnvl", @@ -167,8 +167,8 @@ class TestFlashInferCommFusion(CustomTestCase): for arch in ("pre_sm90", "post_sm10x"): with ( self.subTest(arch=arch), - patch.object(fusion, "is_sm100_supported", return_value=False), - patch.object(fusion, "is_sm90_supported", return_value=False), + override_platform(is_sm100=False), + override_platform(is_sm90=False), ): for args in ( single_node_mnnvl, diff --git a/test/registered/unit/layers/test_minicpm_sparse_metadata.py b/test/registered/unit/layers/test_minicpm_sparse_metadata.py index 80146bec5..de6bc153a 100644 --- a/test/registered/unit/layers/test_minicpm_sparse_metadata.py +++ b/test/registered/unit/layers/test_minicpm_sparse_metadata.py @@ -6,6 +6,7 @@ from unittest.mock import MagicMock, Mock, patch import torch +from sglang.srt.runtime_context import override_platform from sglang.test.ci.ci_register import register_cpu_ci from sglang.test.test_utils import CustomTestCase @@ -96,7 +97,7 @@ def _construct_sparse_backend( with ( get_schedule().override(chunked_prefill_size=chunked_prefill_size), patch.object(backend_module, "MiniCPMHybridConfig", SimpleNamespace), - patch.object(backend_module, "is_blackwell_supported", return_value=blackwell), + override_platform(is_blackwell=blackwell), patch.object( backend_module, "FlashAttentionBackend", @@ -292,7 +293,7 @@ class TestMiniCPMSparseMetadata(CustomTestCase): {"sglang.srt.layers.attention.minicpm.fuse_kernel": fake_fuse_kernel}, ), patch.object(backend_module, "MiniCPMHybridConfig", SimpleNamespace), - patch.object(backend_module, "is_blackwell_supported", return_value=True), + override_platform(is_blackwell=True), patch.object( backend_module, "FlashAttentionBackend", @@ -318,7 +319,7 @@ class TestMiniCPMSparseMetadata(CustomTestCase): model_config.num_attention_heads = 8 with ( patch.object(backend_module, "MiniCPMHybridConfig", SimpleNamespace), - patch.object(backend_module, "is_blackwell_supported", return_value=True), + override_platform(is_blackwell=True), patch.object( backend_module, "FlashAttentionBackend", @@ -337,7 +338,7 @@ class TestMiniCPMSparseMetadata(CustomTestCase): model_runner.server_args.attention_backend = "minicpm_flashattn" with ( patch.object(backend_module, "MiniCPMHybridConfig", SimpleNamespace), - patch.object(backend_module, "is_blackwell_supported", return_value=False), + override_platform(is_blackwell=False), patch.object( backend_module, "FlashAttentionBackend", @@ -397,7 +398,7 @@ class TestMiniCPMSparseMetadata(CustomTestCase): with ( backend_module.envs.SGLANG_MINICPM_DENSE_AS_SPARSE.override(True), patch.object(backend_module, "MiniCPMHybridConfig", SimpleNamespace), - patch.object(backend_module, "is_blackwell_supported", return_value=False), + override_platform(is_blackwell=False), patch.object( backend_module, "FlashAttentionBackend", diff --git a/test/registered/unit/managers/test_mm_process_config.py b/test/registered/unit/managers/test_mm_process_config.py index 5fa803551..cf28ab1b1 100644 --- a/test/registered/unit/managers/test_mm_process_config.py +++ b/test/registered/unit/managers/test_mm_process_config.py @@ -403,8 +403,12 @@ class TestStreamOrderedMmFeaturePool(CustomTestCase): def test_consumer_slot_uses_global_tp_rank(self): from sglang.srt.multimodal.transport.memory_pool import resolve_consumer_rank - parallel = SimpleNamespace(tp_rank=6, attn_tp_rank=2) - with patch("sglang.srt.runtime_context.get_parallel", return_value=parallel): + # State the topology on the context, not by stubbing the accessor: + # `memory_pool` imports `get_parallel` at module scope, so a patch on + # the defining module never reaches the copy doing the reading. + from sglang.srt.runtime_context import get_parallel + + with get_parallel().override(tp_rank=6, attn_tp_rank=2): self.assertEqual(resolve_consumer_rank(8), 6) def test_complete_group_acknowledges_each_consumer_slot(self): diff --git a/test/registered/unit/managers/test_prefill_adder.py b/test/registered/unit/managers/test_prefill_adder.py index 8710ce8f5..6525c3ce9 100644 --- a/test/registered/unit/managers/test_prefill_adder.py +++ b/test/registered/unit/managers/test_prefill_adder.py @@ -13,6 +13,7 @@ from sglang.srt.mem_cache.base_prefix_cache import ( DecLockRefResult, IncLockRefResult, ) +from sglang.srt.runtime_context import get_context from sglang.srt.server_args import ServerArgs, set_global_server_args_for_scheduler from sglang.srt.utils.common import Range from sglang.test.ci.ci_register import register_cpu_ci @@ -79,14 +80,14 @@ class TestPrefillAdder(CustomTestCase): batch.filter_batch.return_value = None return batch - def create_server_args( - self, *, schedule_low_priority_values_first: bool - ) -> MagicMock: - server_args = MagicMock() - server_args.schedule_low_priority_values_first = ( - schedule_low_priority_values_first + def scheduling_order(self, *, schedule_low_priority_values_first: bool): + """State the policy on the context, which is where the scheduler reads + it: `preempt_to_schedule` takes no record to state it on.""" + override = get_context().override_server_args( + schedule_low_priority_values_first=schedule_low_priority_values_first ) - return server_args + override.install() + self.addCleanup(override.restore) def create_mock_req(self, rid, priority, max_new_tokens, output_len=0, wait_time=0): req = MagicMock(spec=Req) @@ -129,9 +130,7 @@ class TestPrefillAdder(CustomTestCase): self.create_mock_req(rid, priority, max_new_tokens) for rid, priority, max_new_tokens in params ] - mock_server_args = self.create_server_args( - schedule_low_priority_values_first=False - ) + self.scheduling_order(schedule_low_priority_values_first=False) running_batch = self.create_running_batch(running_reqs) adder = self.create_adder(running_batch) @@ -144,7 +143,7 @@ class TestPrefillAdder(CustomTestCase): new_req = self.create_mock_req("new1", priority=1, max_new_tokens=49) - success = adder.preempt_to_schedule(new_req, mock_server_args) + success = adder.preempt_to_schedule(new_req) self.assertTrue(success) self.assertIn(running_reqs[0], adder.preempt_list) @@ -161,9 +160,7 @@ class TestPrefillAdder(CustomTestCase): self.create_mock_req(rid, priority, max_new_tokens) for rid, priority, max_new_tokens in params ] - mock_server_args = self.create_server_args( - schedule_low_priority_values_first=True - ) + self.scheduling_order(schedule_low_priority_values_first=True) running_batch = self.create_running_batch(running_reqs) adder = self.create_adder(running_batch) @@ -176,7 +173,7 @@ class TestPrefillAdder(CustomTestCase): new_req = self.create_mock_req("new1", priority=1, max_new_tokens=49) - success = adder.preempt_to_schedule(new_req, mock_server_args) + success = adder.preempt_to_schedule(new_req) self.assertTrue(success) self.assertIn(running_reqs[2], adder.preempt_list) @@ -193,9 +190,7 @@ class TestPrefillAdder(CustomTestCase): self.create_mock_req(rid, priority, max_new_tokens) for rid, priority, max_new_tokens in params ] - mock_server_args = self.create_server_args( - schedule_low_priority_values_first=True - ) + self.scheduling_order(schedule_low_priority_values_first=True) running_batch = self.create_running_batch(running_reqs) adder = self.create_adder(running_batch) @@ -211,7 +206,7 @@ class TestPrefillAdder(CustomTestCase): ) success_by_priority_check = adder.preempt_to_schedule( - new_req_fail_by_priority_check, mock_server_args + new_req_fail_by_priority_check ) self.assertFalse(success_by_priority_check) @@ -219,7 +214,7 @@ class TestPrefillAdder(CustomTestCase): "new2", priority=1, max_new_tokens=110 ) success_by_capacity_check = adder.preempt_to_schedule( - new_req_fail_by_priority_check, mock_server_args + new_req_fail_by_priority_check ) self.assertFalse(success_by_capacity_check) @@ -233,9 +228,7 @@ class TestPrefillAdder(CustomTestCase): self.create_mock_req(rid, priority, max_new_tokens) for rid, priority, max_new_tokens in params ] - mock_server_args = self.create_server_args( - schedule_low_priority_values_first=False - ) + self.scheduling_order(schedule_low_priority_values_first=False) running_batch = self.create_running_batch(running_reqs) adder = self.create_adder(running_batch) @@ -251,7 +244,7 @@ class TestPrefillAdder(CustomTestCase): ) success_by_priority_check = adder.preempt_to_schedule( - new_req_fail_by_priority_check, mock_server_args + new_req_fail_by_priority_check ) self.assertFalse(success_by_priority_check) @@ -259,7 +252,7 @@ class TestPrefillAdder(CustomTestCase): "new2", priority=-1, max_new_tokens=110 ) success_by_capacity_check = adder.preempt_to_schedule( - new_req_fail_by_priority_check, mock_server_args + new_req_fail_by_priority_check ) self.assertFalse(success_by_capacity_check) @@ -273,9 +266,7 @@ class TestPrefillAdder(CustomTestCase): self.create_mock_req(rid, priority, max_new_tokens) for rid, priority, max_new_tokens in params ] - mock_server_args = self.create_server_args( - schedule_low_priority_values_first=False - ) + self.scheduling_order(schedule_low_priority_values_first=False) running_batch = self.create_running_batch(running_reqs) adder = self.create_adder(running_batch) @@ -288,7 +279,7 @@ class TestPrefillAdder(CustomTestCase): first_req = self.create_mock_req( "new_req_prio_1", priority=1, max_new_tokens=49 ) - first_success = adder.preempt_to_schedule(first_req, mock_server_args) + first_success = adder.preempt_to_schedule(first_req) self.assertTrue(first_success) self.assertIn(running_reqs[0], adder.preempt_list) self.assertEqual(adder.rem_total_token_offset, 175) @@ -299,7 +290,7 @@ class TestPrefillAdder(CustomTestCase): second_req = self.create_mock_req( "second_new_req_prio_1", priority=1, max_new_tokens=76 ) - second_success = adder.preempt_to_schedule(second_req, mock_server_args) + second_success = adder.preempt_to_schedule(second_req) self.assertFalse(second_success) self.assertEqual(adder.rem_total_token_offset, 175) @@ -318,9 +309,7 @@ class TestPrefillAdder(CustomTestCase): self.create_mock_req(rid, priority, max_new_tokens) for rid, priority, max_new_tokens in params ] - mock_server_args = self.create_server_args( - schedule_low_priority_values_first=True - ) + self.scheduling_order(schedule_low_priority_values_first=True) running_batch = self.create_running_batch(running_reqs) adder = self.create_adder(running_batch) @@ -333,7 +322,7 @@ class TestPrefillAdder(CustomTestCase): new_req = self.create_mock_req("new1", priority=1, max_new_tokens=75) - success = adder.preempt_to_schedule(new_req, mock_server_args) + success = adder.preempt_to_schedule(new_req) self.assertTrue(success) self.assertIn(running_reqs[2], adder.preempt_list) self.assertEqual( @@ -353,9 +342,7 @@ class TestPrefillAdder(CustomTestCase): self.create_mock_req(rid, priority, max_new_tokens) for rid, priority, max_new_tokens in params ] - mock_server_args = self.create_server_args( - schedule_low_priority_values_first=True - ) + self.scheduling_order(schedule_low_priority_values_first=True) running_batch = self.create_running_batch(running_reqs) adder = self.create_adder(running_batch) @@ -368,7 +355,7 @@ class TestPrefillAdder(CustomTestCase): new_req = self.create_mock_req("new1", priority=1, max_new_tokens=200) - success = adder.preempt_to_schedule(new_req, mock_server_args) + success = adder.preempt_to_schedule(new_req) self.assertTrue(success) self.assertIn(running_reqs[2], adder.preempt_list) self.assertIn(running_reqs[3], adder.preempt_list) diff --git a/test/registered/unit/managers/test_scheduler_hicache_attach.py b/test/registered/unit/managers/test_scheduler_hicache_attach.py index fda42d5aa..d23835705 100644 --- a/test/registered/unit/managers/test_scheduler_hicache_attach.py +++ b/test/registered/unit/managers/test_scheduler_hicache_attach.py @@ -9,6 +9,7 @@ and the published instance stays as the launcher left it. import unittest from types import SimpleNamespace +from sglang.srt.arg_groups.overrides import resolution_result from sglang.srt.managers.io_struct import ( AttachHiCacheStorageReqInput, DetachHiCacheStorageReqInput, @@ -69,7 +70,12 @@ class TestSchedulerHiCacheAttach(CustomTestCase): self.assertIsNone( get_context().resolved_server_args_dict()["hicache_storage_backend"] ) - self.assertEqual(self.server_args.hicache_storage_backend, "file") + # The record is not written any more: the attach is a declaration on + # it and the detach is a bag override (asserted above), so the two are + # meant to differ here. + self.assertEqual( + resolution_result(self.server_args, "hicache_storage_backend"), "file" + ) if __name__ == "__main__": diff --git a/test/registered/unit/models/test_deepseek_v4_amd_fused_mhc.py b/test/registered/unit/models/test_deepseek_v4_amd_fused_mhc.py index 51357a0e7..b5a107a94 100644 --- a/test/registered/unit/models/test_deepseek_v4_amd_fused_mhc.py +++ b/test/registered/unit/models/test_deepseek_v4_amd_fused_mhc.py @@ -3,6 +3,7 @@ from unittest import mock from sglang.srt.environ import envs from sglang.srt.models.deepseek_common.amd import deepseek_v4_fused_mhc +from sglang.srt.runtime_context import override_platform from sglang.test.ci.ci_register import register_cpu_ci register_cpu_ci(est_time=4, suite="base-a-test-cpu") @@ -19,8 +20,8 @@ class TestAmdFusedMhcCrossLayerGating(unittest.TestCase): ): self.assertTrue(deepseek_v4_fused_mhc.is_cross_layer_mhc_fusion_enabled()) - @mock.patch.object(deepseek_v4_fused_mhc, "is_sm120_supported", return_value=True) - def test_sm120_enables_fusion_with_tilelang_pre_disabled(self, _mock_sm120): + @override_platform(is_sm120=True) + def test_sm120_enables_fusion_with_tilelang_pre_disabled(self): # Regression (PR review): consolidating _is_fused_mhc_post_pre_enabled into # this module must preserve the SM120 special case. SM120 disables the # standalone TileLang pre path, but mhc_fused_post_pre dispatches @@ -35,8 +36,8 @@ class TestAmdFusedMhcCrossLayerGating(unittest.TestCase): ): self.assertTrue(deepseek_v4_fused_mhc._is_fused_mhc_post_pre_enabled()) - @mock.patch.object(deepseek_v4_fused_mhc, "is_sm120_supported", return_value=False) - def test_no_sm120_still_requires_tilelang_pre(self, _mock_sm120): + @override_platform(is_sm120=False) + def test_no_sm120_still_requires_tilelang_pre(self): # Negative branch: the (pre OR sm120) clause must not degrade to # always-true. With SM120 unsupported and the pre flag off, fuse+post # alone must not enable the standalone TileLang fused path. @@ -68,11 +69,7 @@ class TestAmdFusedMhcCrossLayerGating(unittest.TestCase): envs.SGLANG_OPT_FUSE_MHC_POST_PRE.override(fuse), envs.SGLANG_OPT_USE_TILELANG_MHC_PRE.override(pre), envs.SGLANG_OPT_USE_TILELANG_MHC_POST.override(post), - mock.patch.object( - deepseek_v4_fused_mhc, - "is_sm120_supported", - return_value=sm120, - ), + override_platform(is_sm120=sm120), ): self.assertEqual( deepseek_v4_fused_mhc._is_fused_mhc_post_pre_enabled(), @@ -83,7 +80,6 @@ class TestAmdFusedMhcCrossLayerGating(unittest.TestCase): @mock.patch.object(deepseek_v4_fused_mhc, "get_bool_env_var", return_value=True) @mock.patch.object(deepseek_v4_fused_mhc, "_is_hip", True) def test_aiter_gfx95_enables_cross_layer_fusion(self, _mock_aiter, _mock_gfx95): - # TileLang flags off: fusion must still enable via the aiter gfx95 path. with ( envs.SGLANG_OPT_FUSE_MHC_POST_PRE.override(False), envs.SGLANG_OPT_USE_TILELANG_MHC_PRE.override(False), diff --git a/test/registered/unit/server_args/test_resolution_reads_no_bag.py b/test/registered/unit/server_args/test_resolution_reads_no_bag.py index 390e6fc94..bdcb053a9 100644 --- a/test/registered/unit/server_args/test_resolution_reads_no_bag.py +++ b/test/registered/unit/server_args/test_resolution_reads_no_bag.py @@ -55,10 +55,9 @@ def _accessor_names(): for node in tree.body if isinstance(node, ast.FunctionDef) and node.name.startswith("get_") } - # The context object itself is not a bag: it exists before anything is - # published, and `declare_late_resolution` calls it deliberately to find - # out whether the record it was handed has been published yet. - return frozenset(names - {"get_context"}) + # Two that are not bags: the context object itself, and the platform facts. + # Both answer before anything is published. + return frozenset(names - {"get_context", "get_platform"}) _BAG_ACCESSORS = _accessor_names() diff --git a/test/registered/unit/server_args/test_resolution_reads_the_declarations.py b/test/registered/unit/server_args/test_resolution_reads_the_declarations.py index dedcde463..ee1283186 100644 --- a/test/registered/unit/server_args/test_resolution_reads_the_declarations.py +++ b/test/registered/unit/server_args/test_resolution_reads_the_declarations.py @@ -471,7 +471,7 @@ class TestResolutionReadsTheDeclarations(CustomTestCase): members = _record_members() # The floor is here to catch the scan collapsing, not to pin the # class's size. - self.assertGreater(len(members), 25, f"only {len(members)} members were found") + self.assertGreater(len(members), 15, f"only {len(members)} members were found") offenders = [] for name, fn in sorted(members.items()): holders = _holders(fn) | {"self"} diff --git a/test/registered/unit/server_args/test_server_args.py b/test/registered/unit/server_args/test_server_args.py index dd60a4ff9..45384d6ff 100644 --- a/test/registered/unit/server_args/test_server_args.py +++ b/test/registered/unit/server_args/test_server_args.py @@ -73,7 +73,12 @@ from sglang.srt.model_executor.cuda_graph_config import ( Phase, PhaseConfig, ) -from sglang.srt.runtime_context import get_context, get_serving +from sglang.srt.runtime_context import ( + describe_kv_events_publisher, + get_context, + get_serving, + override_platform, +) from sglang.srt.server_args import PortArgs, ServerArgs, prepare_server_args from sglang.srt.utils.server_args_config_parser import ConfigArgumentMerger from sglang.test.ci.ci_register import register_cpu_ci @@ -333,8 +338,8 @@ class TestMultimodalFeatureTransport(CustomTestCase): def _set_model_type(server_args, *, is_multimodal): server_args._model_config = SimpleNamespace(is_multimodal=is_multimodal) - @patch("sglang.srt.arg_groups.serving_hook.is_cuda", return_value=True) - def test_cuda_ipc_is_explicit_and_bounded(self, _mock_is_cuda): + @override_platform(is_cuda=True) + def test_cuda_ipc_is_explicit_and_bounded(self): server_args = ServerArgs( model_path="dummy", mm_feature_transport="cuda_ipc", @@ -355,8 +360,8 @@ class TestMultimodalFeatureTransport(CustomTestCase): self.assertIn("base GPU 2", output) self.assertIn("4 tokenizer worker", output) - @patch("sglang.srt.arg_groups.serving_hook.is_cuda", return_value=True) - def test_legacy_keep_flag_maps_to_cuda_ipc(self, _mock_is_cuda): + @override_platform(is_cuda=True) + def test_legacy_keep_flag_maps_to_cuda_ipc(self): server_args = ServerArgs(model_path="dummy", keep_mm_feature_on_device=True) with patch.dict(os.environ, {"SGLANG_USE_CUDA_IPC_TRANSPORT": "0"}): @@ -383,8 +388,8 @@ class TestMultimodalFeatureTransport(CustomTestCase): with self.assertRaisesRegex(ValueError, "conflicts.*cuda_vmm"): handle_multimodal_feature_transport(server_args) - @patch("sglang.srt.arg_groups.serving_hook.is_cuda", return_value=True) - def test_explicit_cpu_overrides_legacy_environment(self, _mock_is_cuda): + @override_platform(is_cuda=True) + def test_explicit_cpu_overrides_legacy_environment(self): server_args = ServerArgs(model_path="dummy", mm_feature_transport="cpu") with patch.dict(os.environ, {"SGLANG_USE_CUDA_IPC_TRANSPORT": "1"}): @@ -409,8 +414,8 @@ class TestMultimodalFeatureTransport(CustomTestCase): ) self.assertFalse(envs.SGLANG_USE_CUDA_IPC_TRANSPORT.get()) - @patch("sglang.srt.arg_groups.serving_hook.is_cuda", return_value=True) - def test_default_transport_is_cpu_for_text_only_model(self, _mock_is_cuda): + @override_platform(is_cuda=True) + def test_default_transport_is_cpu_for_text_only_model(self): server_args = ServerArgs(model_path="dummy") self._set_model_type(server_args, is_multimodal=False) @@ -424,8 +429,8 @@ class TestMultimodalFeatureTransport(CustomTestCase): ) self.assertFalse(envs.SGLANG_USE_CUDA_IPC_TRANSPORT.get()) - @patch("sglang.srt.arg_groups.serving_hook.is_cuda", return_value=True) - def test_default_transport_is_cpu_for_multimodal_model(self, _mock_is_cuda): + @override_platform(is_cuda=True) + def test_default_transport_is_cpu_for_multimodal_model(self): server_args = ServerArgs(model_path="dummy") self._set_model_type(server_args, is_multimodal=True) @@ -443,17 +448,13 @@ class TestMultimodalFeatureTransport(CustomTestCase): @patch( "sglang.srt.arg_groups.serving_hook.is_mnnvl_fabric_device", return_value=True ) - @patch("sglang.srt.arg_groups.serving_hook.is_cuda", return_value=True) + @override_platform(is_cuda=True) @patch( "sglang.srt.model_loader.utils.supports_cuda_vmm_feature_transport", return_value=True, ) def test_default_transport_is_cuda_vmm_for_supported_multinode_mnnvl( - self, - _mock_supports_cuda_vmm, - _mock_is_cuda, - _mock_is_mnnvl, - _mock_path_exists, + self, _mock_supports_cuda_vmm, _mock_is_cuda, _mock_is_mnnvl ): server_args = ServerArgs(model_path="dummy", nnodes=2) self._set_model_type(server_args, is_multimodal=True) @@ -476,17 +477,13 @@ class TestMultimodalFeatureTransport(CustomTestCase): @patch( "sglang.srt.arg_groups.serving_hook.is_mnnvl_fabric_device", return_value=True ) - @patch("sglang.srt.arg_groups.serving_hook.is_cuda", return_value=True) + @override_platform(is_cuda=True) @patch( "sglang.srt.model_loader.utils.supports_cuda_vmm_feature_transport", return_value=False, ) def test_default_transport_is_cpu_for_unsupported_multinode_model( - self, - _mock_supports_cuda_vmm, - _mock_is_cuda, - _mock_is_mnnvl, - _mock_path_exists, + self, _mock_supports_cuda_vmm, _mock_is_cuda, _mock_is_mnnvl ): server_args = ServerArgs(model_path="dummy", nnodes=2) self._set_model_type(server_args, is_multimodal=True) @@ -501,9 +498,9 @@ class TestMultimodalFeatureTransport(CustomTestCase): @patch( "sglang.srt.arg_groups.serving_hook.is_mnnvl_fabric_device", return_value=True ) - @patch("sglang.srt.arg_groups.serving_hook.is_cuda", return_value=True) + @override_platform(is_cuda=True) def test_default_transport_is_cpu_without_imex_channel( - self, _mock_is_cuda, _mock_is_mnnvl, _mock_path_exists + self, _mock_is_cuda, _mock_is_mnnvl ): server_args = ServerArgs(model_path="dummy", nnodes=2) self._set_model_type(server_args, is_multimodal=True) @@ -522,10 +519,8 @@ class TestMultimodalFeatureTransport(CustomTestCase): @patch( "sglang.srt.arg_groups.serving_hook.is_mnnvl_fabric_device", return_value=False ) - @patch("sglang.srt.arg_groups.serving_hook.is_cuda", return_value=True) - def test_default_transport_is_cpu_for_multinode_non_mnnvl( - self, _mock_is_cuda, _mock_is_mnnvl - ): + @override_platform(is_cuda=True) + def test_default_transport_is_cpu_for_multinode_non_mnnvl(self, _mock_is_cuda): server_args = ServerArgs(model_path="dummy", nnodes=2) self._set_model_type(server_args, is_multimodal=True) @@ -538,8 +533,8 @@ class TestMultimodalFeatureTransport(CustomTestCase): ) self.assertFalse(envs.SGLANG_USE_CUDA_IPC_TRANSPORT.get()) - @patch("sglang.srt.arg_groups.serving_hook.is_cuda", return_value=True) - def test_default_transport_is_cpu_for_language_only_model(self, _mock_is_cuda): + @override_platform(is_cuda=True) + def test_default_transport_is_cpu_for_language_only_model(self): server_args = ServerArgs(model_path="dummy", language_only=True) self._set_model_type(server_args, is_multimodal=True) @@ -552,15 +547,15 @@ class TestMultimodalFeatureTransport(CustomTestCase): ) self.assertFalse(envs.SGLANG_USE_CUDA_IPC_TRANSPORT.get()) - @patch("sglang.srt.arg_groups.serving_hook.is_cuda", return_value=False) - def test_cuda_ipc_rejects_non_nvidia_platforms(self, _mock_is_cuda): + @override_platform(is_cuda=False) + def test_cuda_ipc_rejects_non_nvidia_platforms(self): server_args = ServerArgs(model_path="dummy", mm_feature_transport="cuda_ipc") with self.assertRaisesRegex(ValueError, "requires NVIDIA CUDA"): handle_multimodal_feature_transport(server_args) - @patch("sglang.srt.arg_groups.serving_hook.is_cuda", return_value=True) - def test_cuda_ipc_rejects_multi_node(self, _mock_is_cuda): + @override_platform(is_cuda=True) + def test_cuda_ipc_rejects_multi_node(self): server_args = ServerArgs( model_path="dummy", mm_feature_transport="cuda_ipc", nnodes=2 ) @@ -568,8 +563,8 @@ class TestMultimodalFeatureTransport(CustomTestCase): with self.assertRaisesRegex(ValueError, "single node"): handle_multimodal_feature_transport(server_args) - @patch("sglang.srt.arg_groups.serving_hook.is_cuda", return_value=True) - def test_cuda_vmm_is_explicit_and_uses_shared_budget(self, _mock_is_cuda): + @override_platform(is_cuda=True) + def test_cuda_vmm_is_explicit_and_uses_shared_budget(self): server_args = ServerArgs( model_path="dummy", mm_feature_transport="cuda_vmm", @@ -595,15 +590,15 @@ class TestMultimodalFeatureTransport(CustomTestCase): self.assertIn("2 tokenizer worker", output) self.assertIn("falls back to inline CPU", output) - @patch("sglang.srt.arg_groups.serving_hook.is_cuda", return_value=False) - def test_cuda_vmm_rejects_non_nvidia_platforms(self, _mock_is_cuda): + @override_platform(is_cuda=False) + def test_cuda_vmm_rejects_non_nvidia_platforms(self): server_args = ServerArgs(model_path="dummy", mm_feature_transport="cuda_vmm") with self.assertRaisesRegex(ValueError, "requires NVIDIA CUDA"): handle_multimodal_feature_transport(server_args) - @patch("sglang.srt.arg_groups.serving_hook.is_cuda", return_value=True) - def test_cuda_vmm_rejects_rust_server(self, _mock_is_cuda): + @override_platform(is_cuda=True) + def test_cuda_vmm_rejects_rust_server(self): server_args = ServerArgs(model_path="dummy", mm_feature_transport="cuda_vmm") with ( @@ -612,8 +607,8 @@ class TestMultimodalFeatureTransport(CustomTestCase): ): handle_multimodal_feature_transport(server_args) - @patch("sglang.srt.arg_groups.serving_hook.is_cuda", return_value=True) - def test_cuda_vmm_rejects_pipeline_parallelism(self, _mock_is_cuda): + @override_platform(is_cuda=True) + def test_cuda_vmm_rejects_pipeline_parallelism(self): server_args = ServerArgs( model_path="dummy", mm_feature_transport="cuda_vmm", pp_size=2 ) @@ -633,8 +628,8 @@ class TestMambaCacheStochasticRounding(unittest.TestCase): with self.assertRaisesRegex(ValueError, "--mamba-ssm-dtype float16"): handle_mamba_backend(server_args) - @patch("sglang.srt.arg_groups.mamba_hook.is_cuda", return_value=False) - def test_rejects_non_cuda(self, _mock_is_cuda): + @override_platform(is_cuda=False) + def test_rejects_non_cuda(self): server_args = ServerArgs( model_path="dummy", mamba_ssm_dtype="float16", @@ -644,9 +639,9 @@ class TestMambaCacheStochasticRounding(unittest.TestCase): with self.assertRaisesRegex(ValueError, "NVIDIA CUDA"): handle_mamba_backend(server_args) - @patch("sglang.srt.arg_groups.mamba_hook.is_cuda", return_value=True) - @patch("sglang.srt.arg_groups.mamba_hook.is_sm100_supported", return_value=False) - def test_rejects_triton_without_sm100(self, _mock_sm100, _mock_is_cuda): + @override_platform(is_cuda=True) + @override_platform(is_sm100=False) + def test_rejects_triton_without_sm100(self): server_args = ServerArgs( model_path="dummy", mamba_ssm_dtype="float16", @@ -832,8 +827,8 @@ class TestHiSparseDsaBackendPolicy(unittest.TestCase): ) with ( patch("sglang.srt.configs.model_config.is_deepseek_dsa", return_value=True), - patch("sglang.srt.arg_groups.overrides.is_npu", return_value=False), - patch("sglang.srt.arg_groups.overrides.is_xpu", return_value=False), + override_platform(is_npu=False), + override_platform(is_xpu=False), patch("torch.cuda.get_device_capability", return_value=(9, 0)), ): declared = _dsa_split_backend_resolution(view) @@ -846,22 +841,22 @@ class TestHiSparseDsaBackendPolicy(unittest.TestCase): ), } - @patch("sglang.srt.arg_groups.hisparse_hook._is_hip", return_value=False) - def test_hisparse_defaults_to_flashmla_sparse_on_cuda_bfloat16(self, _mock_is_hip): + @override_platform(is_hip=False) + def test_hisparse_defaults_to_flashmla_sparse_on_cuda_bfloat16(self): resolved = self._resolve("bfloat16") self.assertEqual(resolved["dsa_prefill_backend"], "flashmla_sparse") self.assertEqual(resolved["dsa_decode_backend"], "flashmla_sparse") - @patch("sglang.srt.arg_groups.hisparse_hook._is_hip", return_value=False) - def test_hisparse_defaults_to_flashmla_kv_on_cuda_fp8(self, _mock_is_hip): + @override_platform(is_hip=False) + def test_hisparse_defaults_to_flashmla_kv_on_cuda_fp8(self): resolved = self._resolve("fp8_e4m3") self.assertEqual(resolved["dsa_prefill_backend"], "flashmla_kv") self.assertEqual(resolved["dsa_decode_backend"], "flashmla_kv") - @patch("sglang.srt.arg_groups.hisparse_hook._is_hip", return_value=False) - def test_hisparse_accepts_flashinfer_sparse_mla_on_cuda_fp8(self, _mock_is_hip): + @override_platform(is_hip=False) + def test_hisparse_accepts_flashinfer_sparse_mla_on_cuda_fp8(self): """SM120 GLM DSA resolves both DSA backends to flashinfer_sparse_mla, so the fp8 hisparse allow-set must admit it or --enable-hisparse cannot start there at all. The device/arch narrowing happens later, in @@ -877,24 +872,22 @@ class TestHiSparseDsaBackendPolicy(unittest.TestCase): validate_hisparse_dsa_backend(server_args, "dsa_prefill_backend", "prefill") validate_hisparse_dsa_backend(server_args, "dsa_decode_backend", "decode") - @patch("sglang.srt.arg_groups.hisparse_hook._is_hip", return_value=True) - def test_hisparse_defaults_to_tilelang_on_rocm(self, _mock_is_hip): + @override_platform(is_hip=True) + def test_hisparse_defaults_to_tilelang_on_rocm(self): resolved = self._resolve("bfloat16") self.assertEqual(resolved["dsa_prefill_backend"], "tilelang") self.assertEqual(resolved["dsa_decode_backend"], "tilelang") - @patch("sglang.srt.arg_groups.hisparse_hook._is_hip", return_value=True) - def test_hisparse_preserves_rocm_user_backend_and_defaults_missing_side( - self, _mock_is_hip - ): + @override_platform(is_hip=True) + def test_hisparse_preserves_rocm_user_backend_and_defaults_missing_side(self): resolved = self._resolve("bfloat16", dsa_prefill_backend="tilelang") self.assertEqual(resolved["dsa_prefill_backend"], "tilelang") self.assertEqual(resolved["dsa_decode_backend"], "tilelang") - @patch("sglang.srt.arg_groups.hisparse_hook._is_hip", return_value=True) - def test_hisparse_accepts_aiter_backend_on_rocm(self, _mock_is_hip): + @override_platform(is_hip=True) + def test_hisparse_accepts_aiter_backend_on_rocm(self): server_args = ServerArgs( model_path="dummy", enable_hisparse=True, @@ -906,8 +899,8 @@ class TestHiSparseDsaBackendPolicy(unittest.TestCase): validate_hisparse_dsa_backend(server_args, "dsa_prefill_backend", "prefill") validate_hisparse_dsa_backend(server_args, "dsa_decode_backend", "decode") - @patch("sglang.srt.arg_groups.hisparse_hook._is_hip", return_value=True) - def test_hisparse_rejects_cuda_backend_on_rocm(self, _mock_is_hip): + @override_platform(is_hip=True) + def test_hisparse_rejects_cuda_backend_on_rocm(self): server_args = ServerArgs( model_path="dummy", enable_hisparse=True, @@ -918,8 +911,8 @@ class TestHiSparseDsaBackendPolicy(unittest.TestCase): with self.assertRaisesRegex(ValueError, "tilelang"): validate_hisparse_dsa_backend(server_args, "dsa_prefill_backend", "prefill") - @patch("sglang.srt.arg_groups.hisparse_hook._is_hip", return_value=False) - def test_hisparse_rejects_rocm_backend_on_cuda(self, _mock_is_hip): + @override_platform(is_hip=False) + def test_hisparse_rejects_rocm_backend_on_cuda(self): server_args = ServerArgs( model_path="dummy", enable_hisparse=True, @@ -971,15 +964,15 @@ class TestFa4PageSizeAutoForce(CustomTestCase): args.decode_attention_backend = decode args.page_size = page_size # Short-circuit model_config_of(): the fa4 page_size branch only needs - # use_mla_backend() (mocked) and is_sm100_supported() (mocked), not a + # use_mla_backend() (mocked) and override_platform(is_sm100=...), not a # real model_config. Pre-set the attribute so get_model_config returns # early without touching ModelConfig.from_server_args. args._model_config = MagicMock() args._model_config.hf_config.dual_chunk_attention_config = None return args - @patch("sglang.srt.arg_groups.overrides.is_sm100_supported", return_value=True) - def test_combined_attention_backend_fa4_forces_page_size_128(self, _mock_sm100): + @override_platform(is_sm100=True) + def test_combined_attention_backend_fa4_forces_page_size_128(self): # `--attention-backend fa4` (combined): prefill/decode fields stay None. args = self._make_args(attention_backend="fa4") @@ -990,8 +983,8 @@ class TestFa4PageSizeAutoForce(CustomTestCase): self.assertEqual(args.page_size, 1) # the field stays pristine self.assertEqual(resolved_view(args).page_size, 128) - @patch("sglang.srt.arg_groups.overrides.is_sm100_supported", return_value=True) - def test_explicit_prefill_fa4_forces_page_size_128(self, _mock_sm100): + @override_platform(is_sm100=True) + def test_explicit_prefill_fa4_forces_page_size_128(self): # `--prefill-attention-backend fa4`: the previously-covered path. args = self._make_args(attention_backend=None, prefill="fa4", page_size=1) @@ -1969,11 +1962,11 @@ class TestPrefillCudaGraphLoRACompatibility(CustomTestCase): prefill=PhaseConfig(backend=Backend.TC_PIECEWISE) ) with ( - patch("sglang.srt.arg_groups.cuda_graph_hook.is_hip", return_value=False), - patch("sglang.srt.arg_groups.cuda_graph_hook.is_npu", return_value=False), + override_platform(is_hip=False), + override_platform(is_npu=False), patch("sglang.srt.arg_groups.cuda_graph_hook.is_cpu", return_value=False), patch("sglang.srt.arg_groups.cuda_graph_hook.is_mps", return_value=False), - patch("sglang.srt.arg_groups.cuda_graph_hook.is_xpu", return_value=False), + override_platform(is_xpu=False), ): disable_tc_piecewise_cudagraph_if_incompatible(args) @@ -2602,15 +2595,27 @@ class TestGrpcServerArgs(CustomTestCase): ], host="127.0.0.1", ) - # The port the sidecar dials is the resolved one, off the bag. - override = get_context_for_config().override_server_args(grpc_port=50051) + # Every value the sidecar reads is resolved config, so the case states + # them all through the context rather than half here and half in a + # stand-in the readers no longer consult. + override = get_context_for_config().override_server_args( + grpc_port=50051, + sidecar="example.sidecar", + sidecar_args=[ + "--sidecar-shutdown-timeout", + "42", + "--grpc-connections", + "2", + ], + host="127.0.0.1", + ) override.install() self.addCleanup(override.restore) with ( patch("sglang.srt.entrypoints.sidecar.mp.get_context") as get_context, patch("sglang.srt.entrypoints.sidecar.Sidecar") as sidecar_class, ): - start_sidecar(server_args) + start_sidecar() process_kwargs = get_context.return_value.Process.call_args.kwargs self.assertEqual(process_kwargs["name"], "sglang_sidecar_example.sidecar") @@ -2826,11 +2831,11 @@ class TestDcpKvEventContract(CustomTestCase): page_size=64, kv_events_config=self.KV_EVENTS, ) - self.assertEqual(args.describe_kv_events_publisher()["block_size"], 256) + self.assertEqual(describe_kv_events_publisher(args)["block_size"], 256) args = ServerArgs( model_path="dummy", page_size=64, kv_events_config=self.KV_EVENTS ) - self.assertEqual(args.describe_kv_events_publisher()["block_size"], 64) + self.assertEqual(describe_kv_events_publisher(args)["block_size"], 64) def test_kv_event_block_size_widens_a_single_token_page(self): # page_size=1 + DCP is a real deployment shape: the allocator is still diff --git a/test/registered/unit/spec/test_dflash_extra_buffer_lazy.py b/test/registered/unit/spec/test_dflash_extra_buffer_lazy.py index d9ba02159..53238dd38 100644 --- a/test/registered/unit/spec/test_dflash_extra_buffer_lazy.py +++ b/test/registered/unit/spec/test_dflash_extra_buffer_lazy.py @@ -16,6 +16,7 @@ from sglang.test.test_utils import CustomTestCase register_cuda_ci(est_time=10, stage="base-b", runner_config="1-gpu-small") from sglang.srt.model_executor.forward_batch_info import ForwardMode +from sglang.srt.runtime_context import override_platform from sglang.srt.speculative import dflash_info from sglang.srt.speculative.dflash_info import DFlashVerifyInput @@ -29,6 +30,9 @@ def _lazy_view(**overrides): mamba_track_interval=256, page_size=64, chunked_prefill_size=None, + # The cases name a supported arch, so the view states the backend + # rather than the answer. + linear_attn_backend="triton", ) for key, value in overrides.items(): setattr(view, key, value) @@ -39,15 +43,9 @@ class TestValidateMambaExtraBufferLazyDflash(CustomTestCase): """The DFLASH rejection is gone; the neighboring invariants still hold.""" def _validate(self, view): - with mock.patch( - "sglang.srt.arg_groups.overrides.supports_mamba_cache_extra_buffer", - return_value=True, - ), mock.patch( - # Keep the test runnable on CPU-only hosts: the platform assert is - # not what is under test here. - "sglang.srt.arg_groups.mamba_hook.is_cuda", - return_value=True, - ): + # Keep the test runnable on CPU-only hosts: the platform assert is + # not what is under test here. + with override_platform(is_cuda=True): validate_mamba_extra_buffer( view, "Qwen3NextForCausalLM", @@ -81,10 +79,7 @@ class TestValidateMambaExtraBufferLazyDflash(CustomTestCase): def _must_not_be_read(): raise AssertionError("the chunk size was read before page_size resolved") - with mock.patch( - "sglang.srt.arg_groups.overrides.supports_mamba_cache_extra_buffer", - return_value=True, - ), mock.patch("sglang.srt.arg_groups.mamba_hook.is_cuda", return_value=True): + with override_platform(is_cuda=True): validate_mamba_extra_buffer( _lazy_view(page_size=None), "Qwen3NextForCausalLM", diff --git a/test/registered/unit/test_model_overrides.py b/test/registered/unit/test_model_overrides.py index 16068f991..c16b7e76a 100644 --- a/test/registered/unit/test_model_overrides.py +++ b/test/registered/unit/test_model_overrides.py @@ -15,7 +15,6 @@ from types import SimpleNamespace from typing import Optional from unittest.mock import patch -from sglang.srt.arg_groups import attention_hook from sglang.srt.arg_groups import overrides as overrides_module from sglang.srt.arg_groups.arg_utils import A, Arg, resolvable_fields from sglang.srt.arg_groups.overrides import ( @@ -31,6 +30,7 @@ from sglang.srt.runtime_context import ( get_context, get_exec, get_server_args, + override_platform, reset_context, ) from sglang.test.test_utils import CustomTestCase @@ -427,11 +427,7 @@ class TestGoldenModelOverrides(_IsolatedPublish): ) def test_sparse_minicpm_defaults_to_sparse_attention_backend(self): - with patch.object( - overrides_module, - "is_blackwell_supported", - return_value=False, - ): + with override_platform(is_blackwell=False): for architecture in ("MiniCPMForCausalLM", "MiniCPMSALAForCausalLM"): with self.subTest(architecture=architecture): self.assertEqual( @@ -456,20 +452,14 @@ class TestGoldenModelOverrides(_IsolatedPublish): has_lightning_layers=False, ) - with patch.object( - overrides_module, "is_blackwell_supported", return_value=False - ): + with override_platform(is_blackwell=False): overrides = overrides_module._minicpm_sala_overrides(args, config) self.assertTrue(overrides["disable_radix_cache"]) self.assertEqual(overrides["attention_backend"], "minicpm_flashattn") def test_sparse_minicpm_defaults_to_flashinfer_on_blackwell(self): - with patch.object( - overrides_module, - "is_blackwell_supported", - return_value=True, - ): + with override_platform(is_blackwell=True): self.assertEqual( self._minicpm_overrides( "MiniCPMSALAForCausalLM", @@ -531,11 +521,7 @@ class TestGoldenModelOverrides(_IsolatedPublish): )["attention_backend"], "flashinfer", ) - with patch.object( - overrides_module, - "is_blackwell_supported", - return_value=True, - ): + with override_platform(is_blackwell=True): self.assertEqual( self._minicpm_overrides( "MiniCPMSALAForCausalLM", @@ -544,11 +530,7 @@ class TestGoldenModelOverrides(_IsolatedPublish): )["attention_backend"], "fa4", ) - with patch.object( - overrides_module, - "is_blackwell_supported", - return_value=False, - ): + with override_platform(is_blackwell=False): split_overrides = self._minicpm_overrides( "MiniCPMSALAForCausalLM", sparse_attention=True, @@ -625,9 +607,7 @@ class TestGoldenModelOverrides(_IsolatedPublish): """MiniMax-M2 NVFP4 auto must avoid the unsupported plain TRT-LLM path.""" # Every module that asks: the attention handler validates what the # override family picks, and each holds its own import. - with patch.object( - overrides_module, "is_sm100_supported", return_value=True - ), patch.object(attention_hook, "is_sm100_supported", return_value=True): + with override_platform(is_sm100=True), override_platform(is_sm100=True): explicit = self._construct( "MiniMaxM2ForCausalLM", "llama", @@ -666,8 +646,8 @@ class TestGoldenModelOverrides(_IsolatedPublish): # Thor (SM110) and other architectures keep the existing auto behavior. with ( - patch.object(overrides_module, "is_sm100_supported", return_value=False), - patch.object(overrides_module, "is_sm120_supported", return_value=False), + override_platform(is_sm100=False), + override_platform(is_sm120=False), ): non_sm10x = self._construct( "MiniMaxM2ForCausalLM", "llama", quantization="modelopt_fp4" @@ -688,7 +668,7 @@ class TestGoldenModelOverrides(_IsolatedPublish): return SimpleNamespace(**defaults) # Non-SM100: the MoE pin must not fire, so hf_config is never inspected. - with patch.object(overrides_module, "is_sm100_supported", return_value=False): + with override_platform(is_sm100=False): self.assertEqual( _mimo_v2_overrides(_args(speculative_algorithm="EAGLE"), None), {"enable_multi_layer_eagle": True}, @@ -704,7 +684,7 @@ class TestGoldenModelOverrides(_IsolatedPublish): defaults.update(kw) return SimpleNamespace(**defaults) - with patch.object(overrides_module, "is_sm100_supported", return_value=True): + with override_platform(is_sm100=True): with patch.object( overrides_module, "get_quantization_config", return_value="fp8" ): @@ -724,7 +704,7 @@ class TestGoldenModelOverrides(_IsolatedPublish): self.assertEqual(_mimo_v2_overrides(_args(), None), {}) def test_mimo_v2_family_is_registered(self): - with patch.object(overrides_module, "is_sm100_supported", return_value=False): + with override_platform(is_sm100=False): self.assertEqual( collect_model_override_declarations( "MiMoV2FlashForCausalLM", @@ -785,8 +765,8 @@ class TestGoldenModelOverrides(_IsolatedPublish): ) with ( - patch.object(overrides_module, "is_sm100_supported", return_value=True), - patch.object(overrides_module, "is_blackwell_supported", return_value=True), + override_platform(is_sm100=True), + override_platform(is_blackwell=True), ): self.assertEqual( _nemotron_h_overrides(server_args, hf_config), @@ -815,8 +795,8 @@ class TestGoldenModelOverrides(_IsolatedPublish): ) with ( - patch.object(overrides_module, "is_sm100_supported", return_value=True), - patch.object(overrides_module, "is_blackwell_supported", return_value=True), + override_platform(is_sm100=True), + override_platform(is_blackwell=True), ): self.assertEqual( _nemotron_h_overrides(server_args, hf_config), @@ -848,16 +828,8 @@ class TestGoldenModelOverrides(_IsolatedPublish): server_args.speculative_algorithm = "EAGLE" with ( - patch.object( - overrides_module, - "is_blackwell_supported", - return_value=True, - ), - patch.object( - overrides_module, - "is_sm100_supported", - return_value=is_sm100, - ), + override_platform(is_blackwell=True), + override_platform(is_sm100=is_sm100), ): overrides = _nemotron_h_overrides(server_args, hf_config) for key, value in expected.items(): @@ -871,12 +843,8 @@ class TestGoldenModelOverrides(_IsolatedPublish): server_args, hf_config = self._nemotron_h_args(quantized_layers={}) server_args.speculative_algorithm = algorithm with ( - patch.object( - overrides_module, "is_blackwell_supported", return_value=True - ), - patch.object( - overrides_module, "is_sm100_supported", return_value=True - ), + override_platform(is_blackwell=True), + override_platform(is_sm100=True), ): overrides = _nemotron_h_overrides(server_args, hf_config) self.assertEqual(overrides["attention_backend"], "trtllm_mha") @@ -888,8 +856,8 @@ class TestGoldenModelOverrides(_IsolatedPublish): server_args, hf_config = self._nemotron_h_args(quantized_layers={}) server_args.speculative_algorithm = "DFLASH" with ( - patch.object(overrides_module, "is_blackwell_supported", return_value=True), - patch.object(overrides_module, "is_sm100_supported", return_value=True), + override_platform(is_blackwell=True), + override_platform(is_sm100=True), ): overrides = _nemotron_h_overrides(server_args, hf_config) self.assertEqual(overrides["attention_backend"], "trtllm_mha") @@ -905,8 +873,8 @@ class TestGoldenModelOverrides(_IsolatedPublish): server_args.speculative_draft_attention_backend = "flashinfer" with ( - patch.object(overrides_module, "is_blackwell_supported", return_value=True), - patch.object(overrides_module, "is_sm100_supported", return_value=True), + override_platform(is_blackwell=True), + override_platform(is_sm100=True), ): overrides = _nemotron_h_overrides(server_args, hf_config) @@ -923,8 +891,8 @@ class TestGoldenModelOverrides(_IsolatedPublish): server_args.speculative_eagle_topk = 4 with ( - patch.object(overrides_module, "is_blackwell_supported", return_value=True), - patch.object(overrides_module, "is_sm100_supported", return_value=True), + override_platform(is_blackwell=True), + override_platform(is_sm100=True), ): overrides = _nemotron_h_overrides(server_args, hf_config) @@ -939,8 +907,8 @@ class TestGoldenModelOverrides(_IsolatedPublish): server_args, hf_config = self._nemotron_h_args(quantized_layers={}) with ( - patch.object(overrides_module, "is_blackwell_supported", return_value=True), - patch.object(overrides_module, "is_sm100_supported", return_value=False), + override_platform(is_blackwell=True), + override_platform(is_sm100=False), ): self.assertNotIn( "attention_backend", _nemotron_h_overrides(server_args, hf_config) @@ -952,8 +920,8 @@ class TestGoldenModelOverrides(_IsolatedPublish): server_args, hf_config = self._nemotron_h_args(quantized_layers={}) with ( - patch.object(overrides_module, "is_blackwell_supported", return_value=True), - patch.object(overrides_module, "is_sm100_supported", return_value=True), + override_platform(is_blackwell=True), + override_platform(is_sm100=True), ): self.assertEqual( _nemotron_h_overrides(server_args, hf_config)["attention_backend"], @@ -969,8 +937,8 @@ class TestGoldenModelOverrides(_IsolatedPublish): server_args.speculative_draft_attention_backend = "fa3" with ( - patch.object(overrides_module, "is_blackwell_supported", return_value=True), - patch.object(overrides_module, "is_sm100_supported", return_value=True), + override_platform(is_blackwell=True), + override_platform(is_sm100=True), ): overrides = _nemotron_h_overrides(server_args, hf_config) self.assertNotIn("attention_backend", overrides) @@ -1083,7 +1051,7 @@ class TestGoldenModelOverrides(_IsolatedPublish): def test_gpt_oss_xpu_dtype_validation_reads_pristine(self): from sglang.srt.arg_groups.overrides import _gpt_oss_overrides - with patch.object(overrides_module, "is_xpu", return_value=True): + with override_platform(is_xpu=True): with self.assertRaises(NotImplementedError): _gpt_oss_overrides( SimpleNamespace( @@ -1299,12 +1267,12 @@ class TestGoldenModelOverrides(_IsolatedPublish): view = ResolvedView( SimpleNamespace(attention_backend="intel_amx", device="cpu") ) - with patch.object(overrides_module, "cpu_has_amx_support", return_value=False): + with override_platform(has_amx=False): self.assertEqual( _attention_backend_platform_fallbacks(view), {"attention_backend": "torch_native"}, ) - with patch.object(overrides_module, "cpu_has_amx_support", return_value=True): + with override_platform(has_amx=True): self.assertEqual(_attention_backend_platform_fallbacks(view), {}) # dual-chunk config: mismatched explicit backend raises verbatim @@ -1338,20 +1306,20 @@ class TestGoldenModelOverrides(_IsolatedPublish): defaults.update(kw) return ResolvedView(SimpleNamespace(**defaults)) - with patch.object(overrides_module, "is_hip", return_value=True): + with override_platform(is_hip=True): self.assertEqual( _dllm_attention_backend(_view()), {"attention_backend": "triton"} ) self.assertEqual( _dllm_attention_backend(_view(attention_backend="aiter")), {} ) - with patch.object(overrides_module, "is_hip", return_value=False): - with patch.object(overrides_module, "is_npu", return_value=True): + with override_platform(is_hip=False): + with override_platform(is_npu=True): self.assertEqual( _dllm_attention_backend(_view()), {"attention_backend": "ascend"}, ) - with patch.object(overrides_module, "is_npu", return_value=False): + with override_platform(is_npu=False): # cuda graph disabled -> nothing to force self.assertEqual(_dllm_attention_backend(_view()), {}) self.assertEqual( @@ -1366,13 +1334,13 @@ class TestGoldenModelOverrides(_IsolatedPublish): _page_size_default(ResolvedView(SimpleNamespace(page_size=64))), {} ) # default fill on non-HIP/non-MUSA platforms is 1 - with patch.object(overrides_module, "is_hip", return_value=False): - with patch.object(overrides_module, "is_musa", return_value=False): + with override_platform(is_hip=False): + with override_platform(is_musa=False): self.assertEqual( _page_size_default(ResolvedView(SimpleNamespace(page_size=None))), {"page_size": 1}, ) - with patch.object(overrides_module, "is_musa", return_value=True): + with override_platform(is_musa=True): self.assertEqual( _page_size_default(ResolvedView(SimpleNamespace(page_size=None))), {"page_size": 64}, @@ -1477,7 +1445,7 @@ class TestGoldenModelOverrides(_IsolatedPublish): with ( envs.SGLANG_DSV4_FP4_DEQUANT.override(False), - patch.object(overrides_module, "is_sm100_supported", return_value=True), + override_platform(is_sm100=True), ): self.assertEqual( _deepseek_v4_overrides(_args(), hf), @@ -1505,7 +1473,7 @@ class TestGoldenModelOverrides(_IsolatedPublish): # FlashInfer MXFP4 only supports the standard (non-A2A) dispatcher. with ( envs.SGLANG_DSV4_FP4_DEQUANT.override(False), - patch.object(overrides_module, "is_sm100_supported", return_value=True), + override_platform(is_sm100=True), ): self.assertNotIn( "moe_runner_backend", @@ -1514,7 +1482,7 @@ class TestGoldenModelOverrides(_IsolatedPublish): # Runtime FP4-to-FP8 dequantization must retain the generic FP8 runner. with ( envs.SGLANG_DSV4_FP4_DEQUANT.override(True), - patch.object(overrides_module, "is_sm100_supported", return_value=True), + override_platform(is_sm100=True), ): self.assertNotIn( "moe_runner_backend", @@ -1531,7 +1499,7 @@ class TestGoldenModelOverrides(_IsolatedPublish): "moe_runner_backend", _deepseek_v4_overrides(_args(device="npu"), hf), ) - with patch.object(overrides_module, "is_hip", return_value=True): + with override_platform(is_hip=True): self.assertNotIn( "moe_runner_backend", _deepseek_v4_overrides(_args(), hf), @@ -1539,9 +1507,9 @@ class TestGoldenModelOverrides(_IsolatedPublish): # Unsupported NVIDIA architectures keep the generic auto-resolution # path instead of selecting a FlashInfer kernel that cannot launch. with ( - patch.object(overrides_module, "is_sm90_supported", return_value=False), - patch.object(overrides_module, "is_sm100_supported", return_value=False), - patch.object(overrides_module, "is_sm120_supported", return_value=False), + override_platform(is_sm90=False), + override_platform(is_sm100=False), + override_platform(is_sm120=False), ): self.assertNotIn( "moe_runner_backend", @@ -1550,9 +1518,9 @@ class TestGoldenModelOverrides(_IsolatedPublish): # SM120 uses the same model hook; no later pass is needed. with ( envs.SGLANG_DSV4_FP4_DEQUANT.override(False), - patch.object(overrides_module, "is_sm90_supported", return_value=False), - patch.object(overrides_module, "is_sm100_supported", return_value=False), - patch.object(overrides_module, "is_sm120_supported", return_value=True), + override_platform(is_sm90=False), + override_platform(is_sm100=False), + override_platform(is_sm120=True), ): self.assertEqual( _deepseek_v4_overrides(_args(), hf)["moe_runner_backend"], @@ -1605,8 +1573,8 @@ class TestGoldenModelOverrides(_IsolatedPublish): hf = _hf() with ( - patch.object(overrides_module, "is_sm100_supported", return_value=True), - patch.object(overrides_module, "is_blackwell_supported", return_value=True), + override_platform(is_sm100=True), + override_platform(is_blackwell=True), ): # modelopt checkpoint: quant algo resolution + sm100 defaults self.assertEqual( @@ -1625,11 +1593,9 @@ class TestGoldenModelOverrides(_IsolatedPublish): "modelopt_mixed", ) with ( - patch.object(overrides_module, "is_sm100_supported", return_value=False), - patch.object(overrides_module, "is_cuda", return_value=True), - patch.object( - overrides_module, "get_device_capability", return_value=(9, 0) - ), + override_platform(is_sm100=False), + override_platform(is_cuda=True), + override_platform(device_capability=(9, 0)), ): # SM80-SM90 fp4: marlin self.assertEqual( @@ -1651,8 +1617,8 @@ class TestGoldenModelOverrides(_IsolatedPublish): hf_without_quant_cfg = _hf(include_quantization_config=False) with ( - patch.object(overrides_module, "is_sm100_supported", return_value=True), - patch.object(overrides_module, "is_blackwell_supported", return_value=True), + override_platform(is_sm100=True), + override_platform(is_blackwell=True), ): for modelopt_quantization in ("modelopt_fp8", "modelopt_fp4"): with self.subTest(modelopt_quantization=modelopt_quantization): @@ -1718,9 +1684,9 @@ class TestGoldenModelOverrides(_IsolatedPublish): with ( patch("sglang.srt.configs.model_config.is_deepseek_dsa", return_value=True), - patch.object(overrides_module, "is_npu", return_value=False), - patch.object(overrides_module, "is_xpu", return_value=False), - patch.object(overrides_module, "is_hip", return_value=False), + override_platform(is_npu=False), + override_platform(is_xpu=False), + override_platform(is_hip=False), patch("torch.cuda.get_device_capability", return_value=(9, 0)), ): # Hopper FP8 -> flashmla_kv both @@ -1758,9 +1724,9 @@ class TestGoldenModelOverrides(_IsolatedPublish): ) with ( patch("sglang.srt.configs.model_config.is_deepseek_dsa", return_value=True), - patch.object(overrides_module, "is_npu", return_value=False), - patch.object(overrides_module, "is_xpu", return_value=False), - patch.object(overrides_module, "is_hip", return_value=False), + override_platform(is_npu=False), + override_platform(is_xpu=False), + override_platform(is_hip=False), patch("torch.cuda.get_device_capability", return_value=(12, 0)), ): self.assertEqual( @@ -1772,9 +1738,9 @@ class TestGoldenModelOverrides(_IsolatedPublish): ) with ( patch("sglang.srt.configs.model_config.is_deepseek_dsa", return_value=True), - patch.object(overrides_module, "is_npu", return_value=False), - patch.object(overrides_module, "is_xpu", return_value=False), - patch.object(overrides_module, "is_hip", return_value=True), + override_platform(is_npu=False), + override_platform(is_xpu=False), + override_platform(is_hip=True), patch("torch.cuda.get_device_capability", return_value=(9, 4)), ): # ROCm with both unset -> tilelang @@ -1811,8 +1777,8 @@ class TestGoldenModelOverrides(_IsolatedPublish): ) with ( - patch.object(overrides_module, "is_sm90_supported", return_value=True), - patch.object(overrides_module, "is_sm100_supported", return_value=False), + override_platform(is_sm90=True), + override_platform(is_sm100=False), ): self.assertEqual( _flashinfer_allreduce_fusion_auto_enable(_view()), @@ -1897,7 +1863,7 @@ class TestGoldenModelOverrides(_IsolatedPublish): defaults.update(kw) return ResolvedView(SimpleNamespace(**defaults)) - with patch.object(overrides_module, "is_sm100_supported", return_value=True): + with override_platform(is_sm100=True): # decode-only cutedsl: prefill defaults to trtllm_mla self.assertEqual( _cutedsl_prefill_backend_fill(_view()), @@ -1921,7 +1887,7 @@ class TestGoldenModelOverrides(_IsolatedPublish): _cutedsl_prefill_backend_fill(_view(decode_attention_backend=None)), {}, ) - with patch.object(overrides_module, "is_sm100_supported", return_value=False): + with override_platform(is_sm100=False): with self.assertRaises(ValueError): _cutedsl_prefill_backend_fill(_view()) @@ -1971,8 +1937,8 @@ class TestGoldenModelOverrides(_IsolatedPublish): with ( patch("sglang.srt.configs.model_config.is_deepseek_dsa", return_value=True), - patch.object(overrides_module, "is_npu", return_value=False), - patch.object(overrides_module, "is_xpu", return_value=False), + override_platform(is_npu=False), + override_platform(is_xpu=False), ): with patch("torch.cuda.get_device_capability", return_value=(9, 0)): # Hopper: auto -> bfloat16 @@ -2052,7 +2018,7 @@ class TestGoldenModelOverrides(_IsolatedPublish): SimpleNamespace(_model_config=SimpleNamespace(hf_config=hf), **defaults) ) - with patch.object(overrides_module, "is_hip", return_value=True): + with override_platform(is_hip=True): with patch.object( envs.SGLANG_NVFP4_CKPT_FP8_NEXTN_MOE, "get", return_value=False ): @@ -2093,7 +2059,7 @@ class TestGoldenModelOverrides(_IsolatedPublish): with self.assertRaises(ValueError): _deepseek_spec_moe_resolution(_view(ep_size=1)) # the arm is HIP-only - with patch.object(overrides_module, "is_hip", return_value=False): + with override_platform(is_hip=False): self.assertEqual(_deepseek_spec_moe_resolution(_view()), {}) def test_mamba_radix_cache_resolution_pass(self): @@ -2234,9 +2200,7 @@ class TestGoldenModelOverrides(_IsolatedPublish): args._model_config = SimpleNamespace(attention_arch=AttentionArch.MHA) return args - with patch.object( - overrides_module, "is_sm100_supported", return_value=True - ), patch.object( + with override_platform(is_sm100=True), patch.object( overrides_module, "get_default_attn_backend", lambda server_args, **_: server_args.default_backend_for_test, @@ -2278,13 +2242,13 @@ class TestGoldenModelOverrides(_IsolatedPublish): ), {"attention_backend": "trtllm_mha", "page_size": 64}, ) - with patch.object(overrides_module, "is_sm100_supported", return_value=False): + with override_platform(is_sm100=False): self.assertEqual(_qwen3_5_hybrid_overrides(_args("fa3"), None), {}) def test_qwen3vl_page_size(self): from sglang.srt.arg_groups.overrides import _qwen3vl_overrides - with patch.object(overrides_module, "is_hip", return_value=True): + with override_platform(is_hip=True): with patch("sglang.srt.environ.envs.SGLANG_USE_AITER_UNIFIED_ATTN") as e: e.get.return_value = True self.assertEqual( @@ -2306,7 +2270,7 @@ class TestGoldenModelOverrides(_IsolatedPublish): defaults.update(kw) return ResolvedView(SimpleNamespace(**defaults)) - with patch.object(overrides_module, "is_sm100_supported", return_value=True): + with override_platform(is_sm100=True): self.assertEqual( _moe_runner_backend_quant_constraints( _view(quantization="nvfp4_online") @@ -2321,7 +2285,7 @@ class TestGoldenModelOverrides(_IsolatedPublish): _moe_runner_backend_quant_constraints(_view(quantization="mxfp8")), {"moe_runner_backend": "flashinfer_trtllm"}, ) - with patch.object(overrides_module, "is_sm120_supported", return_value=True): + with override_platform(is_sm120=True): self.assertEqual( _moe_runner_backend_quant_constraints( _view(quantization="modelopt_fp4") @@ -2366,7 +2330,7 @@ class TestGoldenModelOverrides(_IsolatedPublish): defaults.update(kw) return SimpleNamespace(**defaults) - with patch("sglang.srt.utils.common.is_sm100_supported", return_value=True): + with override_platform(is_sm100=True): # e4m3 + trtllm_mha + SM100: mode active self.assertTrue(m3_fp8_attn_gemm_enabled(_args())) # fa4 dense backend: mode inactive (no fp8-q GEMM path) @@ -2380,7 +2344,7 @@ class TestGoldenModelOverrides(_IsolatedPublish): # otherwise-active config with envs.SGLANG_DISABLE_M3_FP8_ATTN_GEMM.override(True): self.assertFalse(m3_fp8_attn_gemm_enabled(_args())) - with patch("sglang.srt.utils.common.is_sm100_supported", return_value=False): + with override_platform(is_sm100=False): # non-SM100: mode inactive self.assertFalse(m3_fp8_attn_gemm_enabled(_args())) @@ -2400,8 +2364,8 @@ class TestGoldenModelOverrides(_IsolatedPublish): return ns hf = SimpleNamespace() - with patch.object(overrides_module, "is_hip", return_value=False), patch.object( - overrides_module, "is_sm100_supported", return_value=True + with override_platform(is_hip=False), override_platform( + is_sm100=True ), patch.object(overrides_module, "get_quantization_config", return_value=None): # fp8_e4m3 KV: SM100 backend default flips to trtllm_mha (the only # dense backend with the fp8-q GEMM path); page snaps to 128 @@ -2501,7 +2465,7 @@ class TestGoldenModelOverrides(_IsolatedPublish): # no matching backend: nothing declared self.assertEqual(_mla_backend_page_constraints(_view()), {}) - with patch.object(overrides_module, "is_sm100_supported", return_value=True): + with override_platform(is_sm100=True): self.assertEqual( _fa4_page_constraint( _view( @@ -2565,7 +2529,7 @@ class TestGoldenModelOverrides(_IsolatedPublish): defaults.update(kw) return SimpleNamespace(**defaults) - with patch.object(overrides_module, "is_sm100_supported", return_value=True): + with override_platform(is_sm100=True): self.assertEqual( _llama4_overrides(_args(), None), {"attention_backend": "trtllm_mha"} ) @@ -2599,9 +2563,9 @@ class TestGoldenModelOverrides(_IsolatedPublish): self.assertEqual( _lfm2_overrides(_args(), None), {"attention_backend": "flashinfer"} ) - with patch.object(overrides_module, "is_sm100_supported", return_value=False): + with override_platform(is_sm100=False): self.assertEqual(_minicpm_v4_6_overrides(_args(), None), {}) - with patch.object(overrides_module, "is_sm90_supported", return_value=True): + with override_platform(is_sm90=True): self.assertEqual( _llama4_overrides(_args(), None), {"attention_backend": "fa3"} ) @@ -2609,11 +2573,11 @@ class TestGoldenModelOverrides(_IsolatedPublish): _gemma4_overrides(_args(), None), {"attention_backend": "triton"} ) # Glm4Moe: unconditional tf32 declaration + (sm100) quant/moe absorption - with patch.object(overrides_module, "is_sm100_supported", return_value=False): + with override_platform(is_sm100=False): self.assertEqual( _glm4_moe_overrides(None, None), {"enable_tf32_matmul": True} ) - with patch.object(overrides_module, "is_sm100_supported", return_value=True): + with override_platform(is_sm100=True): self.assertEqual( _glm4_moe_overrides( SimpleNamespace( @@ -2654,7 +2618,7 @@ class TestGoldenModelOverrides(_IsolatedPublish): defaults.update(kw) return ResolvedView(SimpleNamespace(**defaults)) - with patch.object(overrides_module, "is_sm100_supported", return_value=True): + with override_platform(is_sm100=True): with patch.object( overrides_module, "get_quantization_config", return_value="fp8" ): @@ -2670,7 +2634,7 @@ class TestGoldenModelOverrides(_IsolatedPublish): self.assertEqual( _deepseek_moe_quant_resolution(_view(arch="LlamaForCausalLM")), {} ) - with patch.object(overrides_module, "is_sm100_supported", return_value=False): + with override_platform(is_sm100=False): self.assertEqual(_deepseek_moe_quant_resolution(_view()), {}) def test_data_parallelism_and_a2a_passes(self): @@ -2725,15 +2689,15 @@ class TestGoldenModelOverrides(_IsolatedPublish): with patch( "sglang.srt.configs.model_config.is_deepseek_dsa", return_value=True ): - with patch.object(overrides_module, "is_npu", return_value=False): - with patch.object(overrides_module, "is_xpu", return_value=False): - with patch.object(overrides_module, "is_hip", return_value=False): + with override_platform(is_npu=False): + with override_platform(is_xpu=False): + with override_platform(is_hip=False): self.assertEqual( _deepseek_family_overrides(_args(), None), {"attention_backend": "dsa", "page_size": 64}, ) # HIP without the preshuffle path: page 1 - with patch.object(overrides_module, "is_hip", return_value=True): + with override_platform(is_hip=True): with patch( "sglang.srt.layers.attention.dsa.utils.aiter_can_use_preshuffle_paged_mqa", return_value=False, @@ -2746,9 +2710,9 @@ class TestGoldenModelOverrides(_IsolatedPublish): with patch( "sglang.srt.configs.model_config.is_deepseek_dsa", return_value=True ): - with patch.object(overrides_module, "is_npu", return_value=False): - with patch.object(overrides_module, "is_xpu", return_value=False): - with patch.object(overrides_module, "is_hip", return_value=False): + with override_platform(is_npu=False): + with override_platform(is_xpu=False): + with override_platform(is_hip=False): result = _deepseek_family_overrides( _args( enable_prefill_cp=True, @@ -2789,9 +2753,7 @@ class TestGoldenModelOverrides(_IsolatedPublish): with patch( "sglang.srt.configs.model_config.is_deepseek_dsa", return_value=False ): - with patch.object( - overrides_module, "is_sm100_supported", return_value=True - ): + with override_platform(is_sm100=True): self.assertEqual( _deepseek_family_overrides(_args(), None), {"attention_backend": "trtllm_mla"}, @@ -2802,15 +2764,13 @@ class TestGoldenModelOverrides(_IsolatedPublish): ), {}, ) - with patch.object( - overrides_module, "is_sm100_supported", return_value=False - ): + with override_platform(is_sm100=False): self.assertEqual(_deepseek_family_overrides(_args(), None), {}) def test_qwen3_moe_family_quant_absorption(self): from sglang.srt.arg_groups.overrides import _qwen3_moe_family_overrides - with patch.object(overrides_module, "is_sm100_supported", return_value=True): + with override_platform(is_sm100=True): with patch.object( overrides_module, "get_quantization_config", return_value="fp8" ): @@ -2829,7 +2789,7 @@ class TestGoldenModelOverrides(_IsolatedPublish): "moe_runner_backend": "flashinfer_trtllm", }, ) - with patch.object(overrides_module, "is_sm100_supported", return_value=False): + with override_platform(is_sm100=False): self.assertEqual(_qwen3_moe_family_overrides(None, None), {}) def test_step3p_declarations_at_callable_level(self): diff --git a/test/registered/unit/test_platform_address_not_frozen.py b/test/registered/unit/test_platform_address_not_frozen.py new file mode 100644 index 000000000..06b0628dc --- /dev/null +++ b/test/registered/unit/test_platform_address_not_frozen.py @@ -0,0 +1,81 @@ +"""No module-scope name may freeze a platform fact. + +The address exists so `override_platform(...)` reaches every reader at once. +A module-level `_is_sm120 = get_platform().is_sm120` defeats that completely: +the value is read when the module is first imported and never again, so whether +an override is visible depends on import order -- and the line *looks* like it +went through the address, which is worse than the bare probe it replaced. + +Four of these were written during this refactor's own conversion (three in +`fp8_utils`, one in `deepseek_v4_backend`), by substituting the accessor into +lines that were already frozen. Substituting the call is not the conversion; the +conversion is the reader asking at the point of decision. +""" + +from sglang.test.ci.ci_register import register_cpu_ci + +register_cpu_ci(est_time=6, suite="base-a-test-cpu") + +import ast +import pathlib +import unittest + +import sglang +from sglang.test.test_utils import CustomTestCase + +_ROOT = pathlib.Path(next(iter(sglang.__path__))) / "srt" + + +def _frozen_platform_reads(): + """(file, line, name) for each module-scope `x = get_platform().y`.""" + found = [] + for path in sorted(_ROOT.rglob("*.py")): + source = path.read_text(encoding="utf-8-sig") + if "get_platform" not in source: + continue + try: + tree = ast.parse(source) + except SyntaxError: + continue + # Module scope only: inside a function the call runs per invocation, + # which is the shape the address is for. + for node in tree.body: + if not isinstance(node, ast.Assign): + continue + value = node.value + if not ( + isinstance(value, ast.Attribute) + and isinstance(value.value, ast.Call) + and getattr(value.value.func, "id", None) == "get_platform" + ): + continue + for target in node.targets: + if isinstance(target, ast.Name): + rel = path.relative_to(_ROOT).as_posix() + found.append(f"{rel}:{node.lineno} {target.id}") + return found + + +class TestPlatformAddressNotFrozen(CustomTestCase): + def test_the_scan_reaches_the_address(self): + """The premise: `get_platform()` is used somewhere under srt/.""" + users = [ + path + for path in _ROOT.rglob("*.py") + if "get_platform()" in path.read_text(encoding="utf-8-sig") + ] + self.assertGreater(len(users), 20, "the scan found almost no readers") + + def test_no_module_scope_name_freezes_a_platform_fact(self): + frozen = _frozen_platform_reads() + self.assertEqual( + [], + frozen, + "these read a platform fact once at import and keep the answer, so " + "`override_platform(...)` cannot reach them and the result depends " + f"on import order. Ask at the point of decision instead: {frozen}", + ) + + +if __name__ == "__main__": + unittest.main() diff --git a/test/registered/unit/test_platform_context.py b/test/registered/unit/test_platform_context.py new file mode 100644 index 000000000..70003b954 --- /dev/null +++ b/test/registered/unit/test_platform_context.py @@ -0,0 +1,90 @@ +"""A platform fact has one address, and one place to change it. + +The value lives in one object, so `override_platform(...)` reaches every reader +at once, and the facts answer before anything is published. +""" + +import unittest + +from sglang.srt.runtime_context import ( + _PLATFORM_PROBES, + _PLATFORM_VALUES, + get_platform, + override_platform, +) +from sglang.srt.utils import common as _common +from sglang.test.ci.ci_register import register_cpu_ci +from sglang.test.test_utils import CustomTestCase + +register_cpu_ci(est_time=5, suite="base-a-test-cpu") + + +class TestPlatformContext(CustomTestCase): + def test_every_name_maps_to_a_real_probe(self): + """A typo in either table would answer with an AttributeError at runtime.""" + for name, probe in {**_PLATFORM_PROBES, **_PLATFORM_VALUES}.items(): + self.assertTrue( + callable(getattr(_common, probe, None)), + f"{name} maps to {probe!r}, which is not callable in utils.common", + ) + + def test_it_answers_before_anything_is_published(self): + """Unlike a config bag: a launcher asks these before it publishes.""" + platform = get_platform() + for name in _PLATFORM_PROBES: + self.assertIsInstance(getattr(platform, name), bool) + + def test_the_probe_is_what_it_answers_with(self): + platform = get_platform() + for name, probe in {**_PLATFORM_PROBES, **_PLATFORM_VALUES}.items(): + self.assertEqual(getattr(platform, name), getattr(_common, probe)()) + + def test_an_override_is_scoped_and_restores(self): + platform = get_platform() + before = platform.is_sm100 + with override_platform(is_sm100=not before): + self.assertEqual(platform.is_sm100, not before) + self.assertEqual(platform.is_sm100, before) + + def test_overrides_nest(self): + platform = get_platform() + with override_platform(is_sm100=True): + self.assertTrue(platform.is_sm100) + with override_platform(is_sm100=False): + self.assertFalse(platform.is_sm100) + self.assertTrue(platform.is_sm100) + + def test_an_unknown_fact_is_refused_on_both_paths(self): + with self.assertRaises(AttributeError): + get_platform().is_sm999 + with self.assertRaises(ValueError): + override_platform(is_sm999=True).install() + + def test_a_fact_is_not_assigned(self): + """Assignment would move it for one reader, which is the defect.""" + with self.assertRaises(AttributeError): + get_platform().is_sm100 = True + + def test_one_statement_reaches_every_reader(self): + """The point of the address. + + Two modules that both ask the context see the same answer from one + override -- which is what the round-4 failure needed and could not say. + """ + + def module_a(): + return get_platform().is_sm100 + + def module_b(): + return get_platform().is_sm100 + + with override_platform(is_sm100=True): + self.assertTrue(module_a()) + self.assertTrue(module_b()) + with override_platform(is_sm100=False): + self.assertFalse(module_a()) + self.assertFalse(module_b()) + + +if __name__ == "__main__": + unittest.main() diff --git a/test/registered/unit/test_pre_publish_readers.py b/test/registered/unit/test_pre_publish_readers.py new file mode 100644 index 000000000..388e42b13 --- /dev/null +++ b/test/registered/unit/test_pre_publish_readers.py @@ -0,0 +1,134 @@ +"""The launch path runs before `publish`, so nothing on it may ask a config bag. + +`publish` is what projects the bags; every accessor fails closed until then +(`config namespace 'observability' not published`). Most readers live deep in a +runtime path and are safely downstream of it, so converting one to a bag is +normally free. The launcher's own reads are not: everything +`_launch_subprocesses` calls before its `publish` runs with no bags at all, and +`multimodal_gen` calls into the same code with a `ServerArgs` of its own that +never publishes them. + +This is a class no other test in the tree catches: a converted reader is +exercised everywhere by tests that publish first, so it passes unit CI and then +takes the server down on the first real launch -- which is how `configure_logger` +shipped. So the protected set is *derived from the launch path* rather than +listed here: the callees named in `_launch_subprocesses` above its `publish` are +read out of the source, and each is called against a context where nothing has +been published. A conversion of any of them turns this red without a boot. +""" + +from sglang.test.ci.ci_register import register_cpu_ci + +register_cpu_ci(est_time=5, suite="base-a-test-cpu") + +import ast +import logging +import pathlib +import unittest + +import sglang +from sglang.srt.entrypoints.engine import _set_envs_and_config +from sglang.srt.runtime_context import get_observability, reset_context +from sglang.srt.server_args import ServerArgs +from sglang.srt.utils.common import configure_logger +from sglang.test.test_utils import CustomTestCase + +# Callees this guard can drive directly. `_set_envs_and_config` is here too: it +# is the other one that runs on every launch before anything is published. +_EXERCISED = { + "configure_logger": configure_logger, + "_set_envs_and_config": _set_envs_and_config, +} + +# Named by the launcher before `publish`, but none of them reads config out of a +# record. Listed so the set above stays a statement about all of the callees. +_NOT_EXERCISED = { + "load_plugins", + "resolve_auto_parsers", + "snapshot_context", + "resolving_view", +} + + +def _pre_publish_callees(): + """Functions `_launch_subprocesses` calls before it publishes. + + Read from the source so the set cannot go stale: a call added above the + `publish(...)` line joins the protected set on its own. + """ + source = ( + pathlib.Path(next(iter(sglang.__path__))) / "srt" / "entrypoints" / "engine.py" + ).read_text(encoding="utf-8-sig") + tree = ast.parse(source) + launcher = next( + node + for node in ast.walk(tree) + if isinstance(node, ast.FunctionDef) and node.name == "_launch_subprocesses" + ) + publish_line = min( + node.lineno + for node in ast.walk(launcher) + if isinstance(node, ast.Call) and getattr(node.func, "id", None) == "publish" + ) + return { + node.func.id + for node in ast.walk(launcher) + if isinstance(node, ast.Call) + and isinstance(node.func, ast.Name) + and node.lineno < publish_line + } + + +class TestPrePublishReaders(CustomTestCase): + def setUp(self): + self._levels = { + name: logging.getLogger(name).level + for name in (None, "sglang", "httpx", "httpcore") + } + reset_context() + self.addCleanup(self._restore_levels) + self.addCleanup(reset_context) + + def _restore_levels(self): + for name, level in self._levels.items(): + logging.getLogger(name).setLevel(level) + + def test_nothing_is_published_here(self): + """The premise: this fixture really is a pre-publish context.""" + with self.assertRaises(ValueError) as caught: + get_observability() + self.assertIn("not published", str(caught.exception)) + + def test_the_protected_set_is_what_the_launcher_calls(self): + """If the launcher stops calling one of these, or starts calling + something new before publishing, this file has to be looked at.""" + self.assertEqual(set(_EXERCISED) | _NOT_EXERCISED, _pre_publish_callees()) + + def test_none_of_them_asks_a_bag(self): + """Each is called with nothing published. A bag read raises + `config namespace ... not published` -- any other failure is the + callee's own business and does not belong to this guard.""" + server_args = ServerArgs(model_path="dummy", log_level="warning") + for name, call in sorted(_EXERCISED.items()): + with self.subTest(callee=name): + reset_context() + try: + call(server_args) + except Exception as exc: # noqa: BLE001 -- see the docstring + self.assertNotIn( + "not published", + str(exc), + f"{name} runs before publish and asked a config bag", + ) + + def test_configure_logger_reads_the_record_it_was_handed(self): + """The one that shipped broken, pinned by value rather than by not + raising: `multimodal_gen` hands it a ServerArgs that never publishes.""" + configure_logger(ServerArgs(model_path="dummy", log_level="warning")) + self.assertEqual(logging.getLogger().level, logging.WARNING) + configure_logger(ServerArgs(model_path="dummy", log_level="error")) + self.assertEqual(logging.getLogger().level, logging.ERROR) + + +if __name__ == "__main__": + unittest.main() diff --git a/test/registered/unit/test_runtime_context.py b/test/registered/unit/test_runtime_context.py index 36a6b6cbd..cd04d764c 100644 --- a/test/registered/unit/test_runtime_context.py +++ b/test/registered/unit/test_runtime_context.py @@ -26,6 +26,7 @@ from sglang.srt.arg_groups.overrides import ( max_prefill_buffer_tokens as max_prefill_buffer_tokens_of, ) from sglang.srt.arg_groups.overrides import ( + resolution_result, resolved_view, ) from sglang.srt.runtime_context import ( @@ -377,10 +378,12 @@ class TestServerArgsScopedOverride(_IsolatedServerArgs): ) published = override.install() self.assertIs(get_server_args(), published) - self.assertEqual(published.attention_backend, "triton") - self.assertEqual(published.chunked_prefill_size, -1) + # The hook declares; the record keeps the operator's input, so the + # values are read where resolution puts them. + self.assertEqual(resolution_result(published, "attention_backend"), "triton") + self.assertEqual(resolution_result(published, "chunked_prefill_size"), -1) # unnamed fields keep their dataclass defaults - self.assertEqual(published.tp_size, 1) + self.assertEqual(resolution_result(published, "tp_size"), 1) def test_unknown_fields_are_rejected(self): with self.assertRaises(ValueError): @@ -391,7 +394,7 @@ class TestServerArgsScopedOverride(_IsolatedServerArgs): get_context().set_server_args(previous) override = get_context().override_server_args(tp_size=8) override.install() - self.assertEqual(get_server_args().tp_size, 8) + self.assertEqual(get_parallel().tp_size, 8) override.restore() self.assertIs(get_server_args(), previous) @@ -406,9 +409,9 @@ class TestServerArgsScopedOverride(_IsolatedServerArgs): reset_context() with get_context().override_server_args(tp_size=2) as outer: with get_context().override_server_args(tp_size=4): - self.assertEqual(get_server_args().tp_size, 4) + self.assertEqual(get_parallel().tp_size, 4) self.assertIs(get_server_args(), outer) - self.assertEqual(get_server_args().tp_size, 2) + self.assertEqual(get_parallel().tp_size, 2) def test_private_attribute_seeding(self): # Property caches (e.g. _mamba_cache_chunk_size) are seeded through @@ -418,13 +421,35 @@ class TestServerArgsScopedOverride(_IsolatedServerArgs): ) self.assertEqual(mamba_cache_chunk_size_of(published), 64) + def test_an_underscore_field_is_declared_like_any_other(self): + """The split is fields vs not-fields, not the leading underscore. + + `_speculative_draft_quantization_explicitly_set` is a real field + published under `spec`. Seeding it as a raw attribute instead of + declaring it would leave the earlier resolution authoritative, so both + the resolution and the bag would keep answering the pre-override value + while the record said otherwise. + """ + from sglang.srt.arg_groups.overrides import resolution_result + from sglang.srt.runtime_context import get_spec + + name = "_speculative_draft_quantization_explicitly_set" + self.assertIn(name, ServerArgs.__dataclass_fields__) + + published = get_context().override_server_args(**{name: True}).install() + # The record keeps the operator's input, as it does for every other + # field; the override travels as a declaration. + self.assertIsNone(getattr(published, name)) + self.assertIs(resolution_result(published, name), True) + self.assertIs(getattr(get_spec(), name), True) + def test_installed_config_arms_the_strict_guard(self): # The published dummy must behave like a resolved config: bare writes # raise. published = get_context().override_server_args(tp_size=2).install() with self.assertRaises(AttributeError): published.tp_size = 4 - self.assertEqual(published.tp_size, 2) + self.assertEqual(resolution_result(published, "tp_size"), 2) def test_restore_resets_the_capture_seed(self): # install() seeds flags.capture from the published dummy; restore() diff --git a/test/registered/unit/test_server_args_no_instance_mutation_entry.py b/test/registered/unit/test_server_args_no_instance_mutation_entry.py index d6b4e616b..ea6ce5cd3 100644 --- a/test/registered/unit/test_server_args_no_instance_mutation_entry.py +++ b/test/registered/unit/test_server_args_no_instance_mutation_entry.py @@ -98,7 +98,10 @@ class TestNoServerArgsMutationEntry(CustomTestCase): ) def test_late_resolution_refuses_the_published_config(self): - from sglang.srt.arg_groups.overrides import declare_late_resolution + from sglang.srt.arg_groups.overrides import ( + declare_late_resolution, + resolution_result, + ) from sglang.srt.runtime_context import get_context override = get_context().override_server_args(tp_size=2) @@ -107,7 +110,9 @@ class TestNoServerArgsMutationEntry(CustomTestCase): with self.assertRaises(ValueError): declare_late_resolution(published, "test", tp_size=4) - self.assertEqual(published.tp_size, 2) + # The refusal left the resolution alone: the hook's declaration stands, + # and the record still carries the operator's input. + self.assertEqual(resolution_result(published, "tp_size"), 2) if __name__ == "__main__":