Skip to content

Commit b814229

Browse files
authored
[Bugfix] Restrict FlashInfer cuDNN FP8 ViT attention gate to Blackwell (SM 100) (vllm-project#45251)
Signed-off-by: Wentian Byte <3400259131@qq.com>
1 parent 2ec6594 commit b814229

2 files changed

Lines changed: 14 additions & 6 deletions

File tree

vllm/model_executor/layers/attention/mm_encoder_attention.py

Lines changed: 3 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -396,8 +396,9 @@ def _init_fp8_state(self) -> None:
396396
if not is_flashinfer_cudnn_fp8_prefill_attn_supported():
397397
raise ValueError(
398398
"mm_encoder_attn_dtype='fp8' requires the FlashInfer "
399-
"cuDNN backend with cuDNN >= 9.17.1 on a GPU with native "
400-
"FP8 support."
399+
"cuDNN backend with cuDNN >= 9.17.1 on Blackwell (SM 100) "
400+
"or newer. cuDNN's FP8 SDPA path with bf16/fp16 output is "
401+
"not available on Hopper (H100/H200) or earlier."
401402
)
402403

403404
self.fp8_enabled = True

vllm/utils/flashinfer.py

Lines changed: 11 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -934,20 +934,27 @@ def should_use_flashinfer_for_blockscale_fp8_gemm(
934934
return should_use_flashinfer
935935

936936

937-
_MIN_CUDNN_FP8 = 91701 # cuDNN >= 9.17.1 required for FP8 attention
937+
_MIN_CUDNN_FP8 = 91701 # cuDNN >= 9.17.1 required for FP8 ViT attention
938938

939939

940940
@functools.cache
941941
def is_flashinfer_cudnn_fp8_prefill_attn_supported() -> bool:
942942
"""Check if FP8 ViT attention is supported on this platform.
943943
944-
Requires native FP8 hardware support, the FlashInfer cuDNN backend,
944+
Requires Blackwell (SM 100) or newer, the FlashInfer cuDNN backend,
945945
and cuDNN >= 9.17.1.
946+
947+
cuDNN's FP8 SDPA forward path with bf16/fp16 output (used by
948+
``MMEncoderAttention._forward_flashinfer``) gates internally on
949+
``prop.major >= 10``; on Hopper it raises a misleading
950+
``cudnnGraphNotSupportedError: ... cuDNN version 9.13.0 and newer``
951+
even when the installed cuDNN is new enough. See PR #38065 for the
952+
original Blackwell-only design intent.
946953
"""
947954
from vllm.v1.attention.backends.registry import AttentionBackendEnum
948955

949-
# cuDNN SDPA FP8 requires Hopper (SM 90) or newer.
950-
if not current_platform.has_device_capability(90):
956+
# cuDNN SDPA FP8 with bf16/fp16 output requires Blackwell (SM 100) or newer.
957+
if not current_platform.has_device_capability(100):
951958
return False
952959

953960
try:

0 commit comments

Comments
 (0)