[Bug]: failed to set up local/docker vllm with 7900XTX due to skinny_gemms.hip not implemented

[albertokjian]

Your current environment

The output of `python collect_env.py`

INFO 04-28 04:55:30 [__init__.py:239] Automatically detected platform rocm.
Collecting environment information...
PyTorch version: 2.6.0a0+git8d4926e
Is debug build: False
CUDA used to build PyTorch: N/A
ROCM used to build PyTorch: 6.3.42133-1b9c17779

OS: Ubuntu 22.04.5 LTS (x86_64)
GCC version: (Ubuntu 11.4.0-1ubuntu1~22.04) 11.4.0
Clang version: 18.0.0git (https://github.com/RadeonOpenCompute/llvm-project roc-6.3.1 24491 1e0fda770a2079fbd71e4b70974d74f62fd3af10)
CMake version: version 3.31.6
Libc version: glibc-2.35

Python version: 3.12.9 (main, Feb  5 2025, 08:49:00) [GCC 11.4.0] (64-bit runtime)
Python platform: Linux-6.8.0-58-generic-x86_64-with-glibc2.35
Is CUDA available: True
CUDA runtime version: Could not collect
CUDA_MODULE_LOADING set to: LAZY
GPU models and configuration: Radeon RX 7900 XTX (gfx1100)
Nvidia driver version: Could not collect
cuDNN version: Could not collect
HIP runtime version: 6.3.42133
MIOpen runtime version: 3.3.0
Is XNNPACK available: True

CPU:
Architecture:                         x86_64
CPU op-mode(s):                       32-bit, 64-bit
Address sizes:                        46 bits physical, 48 bits virtual
Byte Order:                           Little Endian
CPU(s):                               28
On-line CPU(s) list:                  0-27
Vendor ID:                            GenuineIntel
Model name:                           Intel(R) Core(TM) i7-14700K
CPU family:                           6
Model:                                183
Thread(s) per core:                   2
Core(s) per socket:                   20
Socket(s):                            1
Stepping:                             1
CPU max MHz:                          5600.0000
CPU min MHz:                          800.0000
BogoMIPS:                             6835.20
Flags:                                fpu vme de pse tsc msr pae mce cx8 apic sep mtrr pge mca cmov pat pse36 clflush dts acpi mmx fxsr sse sse2 ss ht tm pbe syscall nx pdpe1gb rdtscp lm constant_tsc art arch_perfmon pebs bts rep_good nopl xtopology nonstop_tsc cpuid aperfmperf tsc_known_freq pni pclmulqdq dtes64 monitor ds_cpl vmx smx est tm2 ssse3 sdbg fma cx16 xtpr pdcm sse4_1 sse4_2 x2apic movbe popcnt tsc_deadline_timer aes xsave avx f16c rdrand lahf_lm abm 3dnowprefetch cpuid_fault ssbd ibrs ibpb stibp ibrs_enhanced tpr_shadow flexpriority ept vpid ept_ad fsgsbase tsc_adjust bmi1 avx2 smep bmi2 erms invpcid rdseed adx smap clflushopt clwb intel_pt sha_ni xsaveopt xsavec xgetbv1 xsaves split_lock_detect user_shstk avx_vnni dtherm ida arat pln pts hwp hwp_notify hwp_act_window hwp_epp hwp_pkg_req hfi vnmi umip pku ospke waitpkg gfni vaes vpclmulqdq tme rdpid movdiri movdir64b fsrm md_clear serialize pconfig arch_lbr ibt flush_l1d arch_capabilities
Virtualization:                       VT-x
L1d cache:                            768 KiB (20 instances)
L1i cache:                            1 MiB (20 instances)
L2 cache:                             28 MiB (11 instances)
L3 cache:                             33 MiB (1 instance)
NUMA node(s):                         1
NUMA node0 CPU(s):                    0-27
Vulnerability Gather data sampling:   Not affected
Vulnerability Itlb multihit:          Not affected
Vulnerability L1tf:                   Not affected
Vulnerability Mds:                    Not affected
Vulnerability Meltdown:               Not affected
Vulnerability Mmio stale data:        Not affected
Vulnerability Reg file data sampling: Mitigation; Clear Register File
Vulnerability Retbleed:               Not affected
Vulnerability Spec rstack overflow:   Not affected
Vulnerability Spec store bypass:      Mitigation; Speculative Store Bypass disabled via prctl
Vulnerability Spectre v1:             Mitigation; usercopy/swapgs barriers and __user pointer sanitization
Vulnerability Spectre v2:             Mitigation; Enhanced / Automatic IBRS; IBPB conditional; RSB filling; PBRSB-eIBRS SW sequence; BHI BHI_DIS_S
Vulnerability Srbds:                  Not affected
Vulnerability Tsx async abort:        Not affected

Versions of relevant libraries:
[pip3] numpy==1.26.4
[pip3] pyzmq==26.4.0
[pip3] torch==2.6.0a0+git8d4926e
[pip3] torchvision==0.19.1a0+6194369
[pip3] transformers==4.51.3
[pip3] triton==3.2.0+gite5be006a
[conda] Could not collect
ROCM Version: 6.3.42133-1b9c17779
Neuron SDK Version: N/A
vLLM Version: 0.8.4
vLLM Build Flags:
CUDA Archs: Not Set; ROCm: Disabled; Neuron: Disabled
GPU Topology:
============================ ROCm System Management Interface ============================
================================ Weight between two GPUs =================================
       GPU0         
GPU0   0            

================================= Hops between two GPUs ==================================
       GPU0         
GPU0   0            

=============================== Link Type between two GPUs ===============================
       GPU0         
GPU0   0            

======================================= Numa Nodes =======================================
GPU[0]		: (Topology) Numa Node: 0
GPU[0]		: (Topology) Numa Affinity: -1
================================== End of ROCm SMI Log ===================================

PYTORCH_ROCM_ARCH=gfx1100;gfx1101;gfx1200;gfx1201
LD_LIBRARY_PATH=/opt/rocm/lib:/usr/local/lib:
NCCL_CUMEM_ENABLE=0
PYTORCH_NVML_BASED_CUDA_CHECK=1
TORCHINDUCTOR_COMPILE_THREADS=1
CUDA_MODULE_LOADING=LAZY

🐛 Describe the bug

I set up vllm with rocm on docker with the following command line using this repo.

DOCKER_BUILDKIT=1 docker build --build-arg BASE_IMAGE="rocm/vllm-dev:navi_base" -f docker/Dockerfile.rocm -t vllm-rocm .

   DOCKER_BUILDKIT=1 docker build --build-arg BASE_IMAGE="rocm/vllm-dev:navi_base" -f docker/Dockerfile.rocm -t vllm-rocm .
[+] Building 960.5s (25/25) FINISHED                                                                                                                                                           docker:default
 => [internal] load build definition from Dockerfile.rocm                                                                                                                                                0.0s
 => => transferring dockerfile: 4.07kB                                                                                                                                                                   0.0s
 => [internal] load metadata for docker.io/rocm/vllm-dev:navi_base                                                                                                                                       0.3s
 => [internal] load .dockerignore                                                                                                                                                                        0.0s
 => => transferring context: 387B                                                                                                                                                                        0.0s
 => [base 1/5] FROM docker.io/rocm/vllm-dev:navi_base@sha256:c7fd65194dbfe67dbc9dd3b5d20b9a8b2c0e4944782cbb6190c57600977b3996                                                                            0.0s
 => [internal] load build context                                                                                                                                                                        0.0s
 => => transferring context: 185.88kB                                                                                                                                                                    0.0s
 => CACHED [base 2/5] RUN apt-get update -q -y && apt-get install -q -y     sqlite3 libsqlite3-dev libfmt-dev libmsgpack-dev libsuitesparse-dev     apt-transport-https ca-certificates wget curl        0.0s
 => CACHED [base 3/5] RUN python3 -m pip install --upgrade pip && pip install setuptools_scm                                                                                                             0.0s
 => CACHED [base 4/5] RUN apt-get purge -y sccache; python3 -m pip uninstall -y sccache; rm -f "$(which sccache)"                                                                                        0.0s
 => CACHED [base 5/5] WORKDIR /app                                                                                                                                                                       0.0s
 => CACHED [fetch_vllm 1/1] ONBUILD COPY ./ vllm/                                                                                                                                                        0.0s
 => [build_vllm 1/1] RUN cd vllm     && python3 -m pip install -r requirements/rocm.txt     && python3 setup.py clean --all      && if [ 0 -eq "1" ]; then python3 tests/build_cython.py build_ext --  893.3s
 => [export_vllm 1/6] COPY --from=build_vllm /app/vllm/dist/*.whl /                                                                                                                                      0.5s 
 => [export_vllm 2/6] COPY --from=build_vllm /app/vllm/requirements /requirements                                                                                                                        0.4s 
 => [export_vllm 3/6] COPY --from=build_vllm /app/vllm/benchmarks /benchmarks                                                                                                                            0.3s 
 => [export_vllm 4/6] COPY --from=build_vllm /app/vllm/tests /tests                                                                                                                                      0.6s 
 => [export_vllm 5/6] COPY --from=build_vllm /app/vllm/examples /examples                                                                                                                                0.2s 
 => [export_vllm 6/6] COPY --from=build_vllm /app/vllm/.buildkite /.buildkite                                                                                                                            0.2s 
 => CACHED [final 1/7] RUN python3 -m pip install --upgrade pip && rm -rf /var/lib/apt/lists/*                                                                                                           0.0s
 => CACHED [final 2/7] RUN case "$(which python3)" in         *"/opt/conda/envs/py_3.9"*)             rm -rf /opt/conda/envs/py_3.9/lib/python3.9/site-packages/numpy-1.20.3.dist-info/;;         *) ;;  0.0s
 => CACHED [final 3/7] RUN python3 -m pip install --upgrade huggingface-hub[cli]                                                                                                                         0.0s
 => CACHED [final 4/7] RUN if [ 1 -eq "1" ]; then     git clone -b nvtx_enabled https://github.com/ROCm/rocmProfileData.git     && cd rocmProfileData/rpd_tracer     && pip install -r requirements.txt  0.0s
 => [final 5/7] RUN --mount=type=bind,from=export_vllm,src=/,target=/install     cd /install     && pip install -U -r requirements/rocm.txt     && pip uninstall -y vllm     && pip install *.whl       52.6s
 => [final 6/7] COPY --from=export_vllm /benchmarks /app/vllm/benchmarks                                                                                                                                 0.3s 
 => [final 7/7] COPY --from=export_vllm /examples /app/vllm/examples                                                                                                                                     0.1s 
 => exporting to image                                                                                                                                                                                   8.5s 
 => => exporting layers                                                                                                                                                                                  8.2s 
 => => writing image sha256:c7f3a7a9bca8fe1e987105c1a8732712b1c52d3c70ed5e12899b150505035a28                                                                                                             0.0s 
 => => naming to docker.io/library/vllm-rocm                                                                                                                                                             0.1s 

Here is the output when running vllm serve "Gensyn/Qwen2.5-0.5B-Instruct" within the docker file.

root@albert-MS-7E06:/app# vllm serve "Gensyn/Qwen2.5-0.5B-Instruct"
DEBUG 04-28 05:02:48 [__init__.py:28] No plugins for group vllm.platform_plugins found.
DEBUG 04-28 05:02:48 [__init__.py:34] Checking if TPU platform is available.
DEBUG 04-28 05:02:48 [__init__.py:44] TPU platform is not available because: No module named 'libtpu'
DEBUG 04-28 05:02:48 [__init__.py:52] Checking if CUDA platform is available.
DEBUG 04-28 05:02:48 [__init__.py:76] Exception happens when checking CUDA platform: NVML Shared Library Not Found
DEBUG 04-28 05:02:48 [__init__.py:93] CUDA platform is not available because: NVML Shared Library Not Found
DEBUG 04-28 05:02:48 [__init__.py:100] Checking if ROCm platform is available.
DEBUG 04-28 05:02:48 [__init__.py:107] Confirmed ROCm platform is available.
DEBUG 04-28 05:02:48 [__init__.py:122] Checking if HPU platform is available.
DEBUG 04-28 05:02:48 [__init__.py:129] HPU platform is not available because habana_frameworks is not found.
DEBUG 04-28 05:02:48 [__init__.py:140] Checking if XPU platform is available.
DEBUG 04-28 05:02:48 [__init__.py:150] XPU platform is not available because: No module named 'intel_extension_for_pytorch'
DEBUG 04-28 05:02:48 [__init__.py:158] Checking if CPU platform is available.
DEBUG 04-28 05:02:48 [__init__.py:180] Checking if Neuron platform is available.
DEBUG 04-28 05:02:48 [__init__.py:187] Neuron platform is not available because: No module named 'transformers_neuronx'
DEBUG 04-28 05:02:48 [__init__.py:100] Checking if ROCm platform is available.
DEBUG 04-28 05:02:48 [__init__.py:107] Confirmed ROCm platform is available.
INFO 04-28 05:02:48 [__init__.py:239] Automatically detected platform rocm.
DEBUG 04-28 05:02:51 [utils.py:135] Setting VLLM_WORKER_MULTIPROC_METHOD to 'spawn'
DEBUG 04-28 05:02:51 [__init__.py:28] No plugins for group vllm.general_plugins found.
INFO 04-28 05:02:51 [api_server.py:1043] vLLM API server version 0.8.4
INFO 04-28 05:02:51 [api_server.py:1044] args: Namespace(subparser='serve', model_tag='Gensyn/Qwen2.5-0.5B-Instruct', config='', host=None, port=8000, uvicorn_log_level='info', disable_uvicorn_access_log=False, allow_credentials=False, allowed_origins=['*'], allowed_methods=['*'], allowed_headers=['*'], api_key=None, lora_modules=None, prompt_adapters=None, chat_template=None, chat_template_content_format='auto', response_role='assistant', ssl_keyfile=None, ssl_certfile=None, ssl_ca_certs=None, enable_ssl_refresh=False, ssl_cert_reqs=0, root_path=None, middleware=[], return_tokens_as_token_ids=False, disable_frontend_multiprocessing=False, enable_request_id_headers=False, enable_auto_tool_choice=False, tool_call_parser=None, tool_parser_plugin='', model='Gensyn/Qwen2.5-0.5B-Instruct', task='auto', tokenizer=None, hf_config_path=None, skip_tokenizer_init=False, revision=None, code_revision=None, tokenizer_revision=None, tokenizer_mode='auto', trust_remote_code=False, allowed_local_media_path=None, load_format='auto', download_dir=None, model_loader_extra_config={}, use_tqdm_on_load=True, config_format=<ConfigFormat.AUTO: 'auto'>, dtype='auto', max_model_len=None, guided_decoding_backend='auto', reasoning_parser=None, logits_processor_pattern=None, model_impl='auto', distributed_executor_backend=None, pipeline_parallel_size=1, tensor_parallel_size=1, data_parallel_size=1, enable_expert_parallel=False, max_parallel_loading_workers=None, ray_workers_use_nsight=False, disable_custom_all_reduce=False, block_size=None, gpu_memory_utilization=0.9, swap_space=4, kv_cache_dtype='auto', num_gpu_blocks_override=None, enable_prefix_caching=None, prefix_caching_hash_algo='builtin', cpu_offload_gb=0, calculate_kv_scales=False, disable_sliding_window=False, use_v2_block_manager=True, seed=None, max_logprobs=20, disable_log_stats=False, quantization=None, rope_scaling=None, rope_theta=None, hf_token=None, hf_overrides=None, enforce_eager=False, max_seq_len_to_capture=8192, tokenizer_pool_size=0, tokenizer_pool_type='ray', tokenizer_pool_extra_config={}, limit_mm_per_prompt={}, mm_processor_kwargs=None, disable_mm_preprocessor_cache=False, enable_lora=None, enable_lora_bias=False, max_loras=1, max_lora_rank=16, lora_extra_vocab_size=256, lora_dtype='auto', long_lora_scaling_factors=None, max_cpu_loras=None, fully_sharded_loras=False, enable_prompt_adapter=None, max_prompt_adapters=1, max_prompt_adapter_token=0, device='auto', speculative_config=None, ignore_patterns=[], served_model_name=None, qlora_adapter_name_or_path=None, show_hidden_metrics_for_version=None, otlp_traces_endpoint=None, collect_detailed_traces=None, disable_async_output_proc=False, max_num_batched_tokens=None, max_num_seqs=None, max_num_partial_prefills=1, max_long_partial_prefills=1, long_prefill_token_threshold=0, num_lookahead_slots=0, scheduler_delay_factor=0.0, preemption_mode=None, num_scheduler_steps=1, multi_step_stream_outputs=True, scheduling_policy='fcfs', enable_chunked_prefill=None, disable_chunked_mm_input=False, scheduler_cls='vllm.core.scheduler.Scheduler', override_neuron_config=None, override_pooler_config=None, compilation_config=None, kv_transfer_config=None, worker_cls='auto', worker_extension_cls='', generation_config='auto', override_generation_config=None, enable_sleep_mode=False, additional_config=None, enable_reasoning=False, disable_cascade_attn=False, disable_log_requests=False, max_log_len=None, disable_fastapi_docs=False, enable_prompt_tokens_details=False, enable_server_load_tracking=False, dispatch_function=<function ServeSubcommand.cmd at 0x7efb57c71620>)
INFO 04-28 05:02:57 [config.py:717] This model supports multiple tasks: {'embed', 'score', 'reward', 'generate', 'classify'}. Defaulting to 'generate'.
INFO 04-28 05:02:57 [arg_utils.py:1682] rocm is experimental on VLLM_USE_V1=1. Falling back to V0 Engine.
INFO 04-28 05:02:57 [config.py:1784] Disabled the custom all-reduce kernel because it is not supported on current platform.
INFO 04-28 05:02:57 [config.py:1755] Disabled the custom all-reduce kernel because it is not working correctly when using two AMD Navi GPUs.
DEBUG 04-28 05:02:57 [api_server.py:223] Multiprocessing frontend to use ipc:///tmp/1361f47e-1dd3-4356-847e-2b287e8dcd01 for IPC Path.
INFO 04-28 05:02:57 [api_server.py:246] Started engine process with PID 677
DEBUG 04-28 05:02:59 [__init__.py:28] No plugins for group vllm.platform_plugins found.
DEBUG 04-28 05:02:59 [__init__.py:34] Checking if TPU platform is available.
DEBUG 04-28 05:02:59 [__init__.py:44] TPU platform is not available because: No module named 'libtpu'
DEBUG 04-28 05:02:59 [__init__.py:52] Checking if CUDA platform is available.
DEBUG 04-28 05:02:59 [__init__.py:76] Exception happens when checking CUDA platform: NVML Shared Library Not Found
DEBUG 04-28 05:02:59 [__init__.py:93] CUDA platform is not available because: NVML Shared Library Not Found
DEBUG 04-28 05:02:59 [__init__.py:100] Checking if ROCm platform is available.
DEBUG 04-28 05:02:59 [__init__.py:107] Confirmed ROCm platform is available.
DEBUG 04-28 05:02:59 [__init__.py:122] Checking if HPU platform is available.
DEBUG 04-28 05:02:59 [__init__.py:129] HPU platform is not available because habana_frameworks is not found.
DEBUG 04-28 05:02:59 [__init__.py:140] Checking if XPU platform is available.
DEBUG 04-28 05:02:59 [__init__.py:150] XPU platform is not available because: No module named 'intel_extension_for_pytorch'
DEBUG 04-28 05:02:59 [__init__.py:158] Checking if CPU platform is available.
DEBUG 04-28 05:02:59 [__init__.py:180] Checking if Neuron platform is available.
DEBUG 04-28 05:02:59 [__init__.py:187] Neuron platform is not available because: No module named 'transformers_neuronx'
DEBUG 04-28 05:02:59 [__init__.py:100] Checking if ROCm platform is available.
DEBUG 04-28 05:02:59 [__init__.py:107] Confirmed ROCm platform is available.
INFO 04-28 05:02:59 [__init__.py:239] Automatically detected platform rocm.
DEBUG 04-28 05:03:02 [__init__.py:28] No plugins for group vllm.general_plugins found.
INFO 04-28 05:03:02 [llm_engine.py:242] Initializing a V0 LLM engine (v0.8.4) with config: model='Gensyn/Qwen2.5-0.5B-Instruct', speculative_config=None, tokenizer='Gensyn/Qwen2.5-0.5B-Instruct', skip_tokenizer_init=False, tokenizer_mode=auto, revision=None, override_neuron_config=None, tokenizer_revision=None, trust_remote_code=False, dtype=torch.bfloat16, max_seq_len=32768, download_dir=None, load_format=LoadFormat.AUTO, tensor_parallel_size=1, pipeline_parallel_size=1, disable_custom_all_reduce=True, quantization=None, enforce_eager=False, kv_cache_dtype=auto,  device_config=cuda, decoding_config=DecodingConfig(guided_decoding_backend='auto', reasoning_backend=None), observability_config=ObservabilityConfig(show_hidden_metrics=False, otlp_traces_endpoint=None, collect_model_forward_time=False, collect_model_execute_time=False), seed=None, served_model_name=Gensyn/Qwen2.5-0.5B-Instruct, num_scheduler_steps=1, multi_step_stream_outputs=True, enable_prefix_caching=None, chunked_prefill_enabled=False, use_async_output_proc=True, disable_mm_preprocessor_cache=False, mm_processor_kwargs=None, pooler_config=None, compilation_config={"splitting_ops":[],"compile_sizes":[],"cudagraph_capture_sizes":[256,248,240,232,224,216,208,200,192,184,176,168,160,152,144,136,128,120,112,104,96,88,80,72,64,56,48,40,32,24,16,8,4,2,1],"max_capture_size":256}, use_cached_outputs=True, 
INFO 04-28 05:03:02 [rocm.py:186] None is not supported in AMD GPUs.
INFO 04-28 05:03:02 [rocm.py:187] Using ROCmFlashAttention backend.
DEBUG 04-28 05:03:02 [config.py:4034] enabled custom ops: Counter()
DEBUG 04-28 05:03:02 [config.py:4036] disabled custom ops: Counter()
DEBUG 04-28 05:03:02 [parallel_state.py:809] world_size=1 rank=0 local_rank=0 distributed_init_method=tcp://192.168.0.19:33625 backend=nccl
INFO 04-28 05:03:02 [parallel_state.py:946] rank 0 in world size 1 is assigned as DP rank 0, PP rank 0, TP rank 0
DEBUG 04-28 05:03:02 [config.py:4034] enabled custom ops: Counter()
DEBUG 04-28 05:03:02 [config.py:4036] disabled custom ops: Counter()
INFO 04-28 05:03:02 [model_runner.py:1120] Starting to load model Gensyn/Qwen2.5-0.5B-Instruct...
WARNING 04-28 05:03:02 [rocm.py:288] Model architecture 'Qwen2ForCausalLM' is partially supported by ROCm: Sliding window attention (SWA) is not yet supported in Triton flash attention. For half-precision SWA support, please use CK flash attention by setting `VLLM_USE_TRITON_FLASH_ATTN=0`
DEBUG 04-28 05:03:02 [rocm_flash_attn.py:553] Using Triton FA in ROCmBackend
DEBUG 04-28 05:03:02 [rocm_flash_attn.py:553] Using Triton FA in ROCmBackend
DEBUG 04-28 05:03:02 [rocm_flash_attn.py:553] Using Triton FA in ROCmBackend
DEBUG 04-28 05:03:02 [rocm_flash_attn.py:553] Using Triton FA in ROCmBackend
DEBUG 04-28 05:03:02 [rocm_flash_attn.py:553] Using Triton FA in ROCmBackend
DEBUG 04-28 05:03:02 [rocm_flash_attn.py:553] Using Triton FA in ROCmBackend
DEBUG 04-28 05:03:02 [rocm_flash_attn.py:553] Using Triton FA in ROCmBackend
DEBUG 04-28 05:03:02 [rocm_flash_attn.py:553] Using Triton FA in ROCmBackend
DEBUG 04-28 05:03:02 [rocm_flash_attn.py:553] Using Triton FA in ROCmBackend
DEBUG 04-28 05:03:02 [rocm_flash_attn.py:553] Using Triton FA in ROCmBackend
DEBUG 04-28 05:03:02 [rocm_flash_attn.py:553] Using Triton FA in ROCmBackend
DEBUG 04-28 05:03:02 [rocm_flash_attn.py:553] Using Triton FA in ROCmBackend
DEBUG 04-28 05:03:02 [rocm_flash_attn.py:553] Using Triton FA in ROCmBackend
DEBUG 04-28 05:03:02 [rocm_flash_attn.py:553] Using Triton FA in ROCmBackend
DEBUG 04-28 05:03:02 [rocm_flash_attn.py:553] Using Triton FA in ROCmBackend
DEBUG 04-28 05:03:02 [rocm_flash_attn.py:553] Using Triton FA in ROCmBackend
DEBUG 04-28 05:03:02 [rocm_flash_attn.py:553] Using Triton FA in ROCmBackend
DEBUG 04-28 05:03:02 [rocm_flash_attn.py:553] Using Triton FA in ROCmBackend
DEBUG 04-28 05:03:02 [rocm_flash_attn.py:553] Using Triton FA in ROCmBackend
DEBUG 04-28 05:03:02 [rocm_flash_attn.py:553] Using Triton FA in ROCmBackend
DEBUG 04-28 05:03:02 [rocm_flash_attn.py:553] Using Triton FA in ROCmBackend
DEBUG 04-28 05:03:02 [rocm_flash_attn.py:553] Using Triton FA in ROCmBackend
DEBUG 04-28 05:03:02 [rocm_flash_attn.py:553] Using Triton FA in ROCmBackend
DEBUG 04-28 05:03:02 [rocm_flash_attn.py:553] Using Triton FA in ROCmBackend
DEBUG 04-28 05:03:02 [config.py:4034] enabled custom ops: Counter({'rms_norm': 49, 'silu_and_mul': 24, 'rotary_embedding': 1})
DEBUG 04-28 05:03:02 [config.py:4036] disabled custom ops: Counter()
INFO 04-28 05:03:03 [weight_utils.py:265] Using model weights format ['*.safetensors']
INFO 04-28 05:03:03 [weight_utils.py:315] No model.safetensors.index.json found in remote.
Loading safetensors checkpoint shards:   0% Completed | 0/1 [00:00<?, ?it/s]
Loading safetensors checkpoint shards: 100% Completed | 1/1 [00:00<00:00,  2.00it/s]
Loading safetensors checkpoint shards: 100% Completed | 1/1 [00:00<00:00,  2.00it/s]

INFO 04-28 05:03:03 [loader.py:458] Loading weights took 0.54 seconds
INFO 04-28 05:03:04 [model_runner.py:1156] Model loading took 1.1445 GiB and 1.173828 seconds
INFO 04-28 05:03:05 [worker.py:315] Memory profiling takes 1.81 seconds
INFO 04-28 05:03:05 [worker.py:315] the current vLLM instance can use total_gpu_memory (23.98GiB) x gpu_memory_utilization (0.90) = 21.59GiB
INFO 04-28 05:03:05 [worker.py:315] model weights take 1.14GiB; non_torch_memory takes 0.27GiB; PyTorch activation peak memory takes 1.43GiB; the rest of the memory reserved for KV Cache is 18.74GiB.
INFO 04-28 05:03:06 [executor_base.py:112] # rocm blocks: 102318, # CPU blocks: 21845
INFO 04-28 05:03:06 [executor_base.py:117] Maximum concurrency for 32768 tokens per request: 49.96x
DEBUG 04-28 05:03:08 [client.py:193] Waiting for output from MQLLMEngine.
INFO 04-28 05:03:08 [model_runner.py:1466] Capturing cudagraphs for decoding. This may lead to unexpected consequences if the model is not static. To run the model in eager mode, set 'enforce_eager=True' or use '--enforce-eager' in the CLI. If out-of-memory error occurs during cudagraph capture, consider decreasing `gpu_memory_utilization` or switching to eager mode. You can also reduce the `max_num_seqs` as needed to decrease memory usage.
Capturing CUDA graph shapes:  94%|████████████████████████████████████████████████████████████████████████████████████████████████████████████████████████████████████        | 33/35 [00:06<00:00,  4.90it/s]/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
/app/vllm/build/temp.linux-x86_64-cpython-312/csrc/rocm/skinny_gemms.hip:497: void wvSplitK_hf_sml_(const int, const int, const scalar_t *, const scalar_t *__restrict, scalar_t *, const int, const int) [scalar_t = __hip_bfloat16, THRDS = 64, YTILE = 2, WvPrGrp = 16, A_CHUNK = 8, UNRL = 2, N = 2]: Device-side assertion `false' failed.
:0:rocdevice.cpp            :3018: 200246735247d us:  Callback: Queue 0x76005c600000 aborting with error : HSA_STATUS_ERROR_EXCEPTION: An HSAIL operation resulted in a hardware exception. code: 0x1016
DEBUG 04-28 05:03:18 [client.py:193] Waiting for output from MQLLMEngine.
DEBUG 04-28 05:03:28 [client.py:260] Shutting down MQLLMEngineClient output handler.
Traceback (most recent call last):
  File "/usr/local/bin/vllm", line 8, in <module>
    sys.exit(main())
             ^^^^^^
  File "/usr/local/lib/python3.12/dist-packages/vllm/entrypoints/cli/main.py", line 53, in main
    args.dispatch_function(args)
  File "/usr/local/lib/python3.12/dist-packages/vllm/entrypoints/cli/serve.py", line 27, in cmd
    uvloop.run(run_server(args))
  File "/usr/local/lib/python3.12/dist-packages/uvloop/__init__.py", line 109, in run
    return __asyncio.run(
           ^^^^^^^^^^^^^^
  File "/usr/lib/python3.12/asyncio/runners.py", line 195, in run
    return runner.run(main)
           ^^^^^^^^^^^^^^^^
  File "/usr/lib/python3.12/asyncio/runners.py", line 118, in run
    return self._loop.run_until_complete(task)
           ^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^
  File "uvloop/loop.pyx", line 1518, in uvloop.loop.Loop.run_until_complete
  File "/usr/local/lib/python3.12/dist-packages/uvloop/__init__.py", line 61, in wrapper
    return await main
           ^^^^^^^^^^
  File "/usr/local/lib/python3.12/dist-packages/vllm/entrypoints/openai/api_server.py", line 1078, in run_server
    async with build_async_engine_client(args) as engine_client:
               ^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^
  File "/usr/lib/python3.12/contextlib.py", line 210, in __aenter__
    return await anext(self.gen)
           ^^^^^^^^^^^^^^^^^^^^^
  File "/usr/local/lib/python3.12/dist-packages/vllm/entrypoints/openai/api_server.py", line 146, in build_async_engine_client
    async with build_async_engine_client_from_engine_args(
               ^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^
  File "/usr/lib/python3.12/contextlib.py", line 210, in __aenter__
    return await anext(self.gen)
           ^^^^^^^^^^^^^^^^^^^^^
  File "/usr/local/lib/python3.12/dist-packages/vllm/entrypoints/openai/api_server.py", line 269, in build_async_engine_client_from_engine_args
    raise RuntimeError(
RuntimeError: Engine process failed to start. See stack trace for the root cause.
root@albert-MS-7E06:/app# /usr/lib/python3.12/multiprocessing/resource_tracker.py:255: UserWarning: resource_tracker: There appear to be 1 leaked semaphore objects to clean up at shutdown
  warnings.warn('resource_tracker: There appear to be %d '

I have also tried the set up instructions for building wheel from source to no avail. The particular failure above is probably due to the following code under csrc\rocm\skinny_gemms.cu not having NAVI support.

#else   // !defined(__HIP__MI300_MI250__) TODO: Add NAVI support
template <typename scalar_t, int THRDS, int YTILE, int WvPrGrp, int A_CHUNK,
          int UNRL, int N>
__global__ void wvSplitK_hf_sml_(const int K, const int M, const scalar_t* B,
                                 const scalar_t* __restrict__ A, scalar_t* C,
                                 const int _WvPrGrp, const int CuCount) {
  UNREACHABLE_CODE
}

I am able to run the model via transformer within the docker.

root@albert-MS-7E06:/app# python
Python 3.12.9 (main, Feb  5 2025, 08:49:00) [GCC 11.4.0] on linux
Type "help", "copyright", "credits" or "license" for more information.
>>> # Use a pipeline as a high-level helper
>>> from transformers import pipeline
>>> 
>>> messages = [
...     {"role": "user", "content": "Who are you?"},
... ]
>>> pipe = pipeline("text-generation", model="Gensyn/Qwen2.5-0.5B-Instruct")
Sliding Window Attention is enabled but not implemented for `sdpa`; unexpected results may be encountered.
Device set to use cuda:0
>>> pipe(messages)
[{'generated_text': [{'role': 'user', 'content': 'Who are you?'}, {'role': 'assistant', 'content': 'I am Qwen, an artificial intelligence language model created by Alibaba Cloud. I am designed to assist'}]}]

Before submitting a new issue...

• Make sure you already searched for relevant issues, and asked the chatbot living at the bottom right corner of the documentation page, which can answer lots of frequently asked questions.

[gshtras]

Thank you for reporting. This should be fixed by vllm-project#17011 once we pull it in.
Sadly it currently has a bug around the condition gating Radeon, but it is identified and will be fixed soon

[albertokjian]

Thanks for your quick response! I was having trouble running vllm-project's most up-to-date version, and was hoping that this repo would work. I will wait for your patch then.