[gfx1201/gfx1151] Collecting MIOpen and hipBLASLt logs

[jammm]

Are you facing slow performance when running your models using ComfyUI/SD WebUI or any pytorch program using your Radeon 9070XT, AI Pro R9700, or Strix Halo (Radeon 8060S) ? Then we need your help! Please provide us performance logs when running your models. It will help us tune our libraries for better performance on your models.

Please set the following environment variables depending on your OS:

Windows (powershell)

$env:MIOPEN_ENABLE_LOGGING=1
$env:MIOPEN_ENABLE_LOGGING_CMD=1
$env:HIPBLASLT_LOG_MASK=32
$env:HIPBLASLT_LOG_FILE="your_model_hipblaslt_log.txt"
$env:TORCH_BLAS_PREFER_HIPBLASLT=1
$env:COMFYUI_ENABLE_MIOPEN=1 

Linux:

export MIOPEN_ENABLE_LOGGING=1
export MIOPEN_ENABLE_LOGGING_CMD=1
export HIPBLASLT_LOG_MASK=32
export HIPBLASLT_LOG_FILE="your_model_hipblaslt_log.txt"
export TORCH_BLAS_PREFER_HIPBLASLT=1
export COMFYUI_ENABLE_MIOPEN=1 

NOTE: hipBLASLt logs every call of every kernel and if you run ComfyUI for more images or some LLM with more prompts, log file can grow to several hundred MBs.
To extract unique GEMM shapes we can sort our log file before uploading. Please use below commands to get only unique ones sorted by number of occurrences which can help us a lot to focus on the most used GEMM kernels.

Windows (powershell)

Get-Content your_model_hipblaslt_log.txt | Group-Object | Sort-Object Count -Descending | Select-Object Count, Name > your_model_sorted_hipblaslt_log.txt

Linux:

sort your_model_hipblaslt_log.txt | uniq -c | sort -nr > your_model_sorted_hipblaslt_log.txt

IMPORTANT: Please make sure you're using the latest ComfyUI master. If you're not using the latest ComfyUI master, please remove/comment out torch.backends.cudnn.enabled = False from comfy/model_management.py within your ComfyUI folder in order to enable MIOpen so the logs can be collected. If you are using the latest ComfyUI from master branch as of December 20th, then you do not need to make this change.

NOTE: Make sure to delete your .miopen folder between runs - e.g., from C:\Users\<username>\.miopen.

An alternative is to add the following line right after import torch in your pytorch scripts:

type(torch.backends.cudnn).enabled = property(lambda self: True, lambda self, v: None)

Then run your workload as usual, for example, SD WebUI, your pytorch scripts, or ComfyUI, using (command) > miopen_output_logs.txt 2>&1 on the same terminal you used to run the above commands.

Then share with us the output logs from hipBLASLt (your_model_hipblaslt_log.txt) and output+MIOpen logs miopen_output_logs.txt from the terminal output in the comments below.

[astrelsky]

Will you take gfx1150 or does it need to be gfx1151?

I think the logs you want to collect are written to stderr?

To collect both stdout and stderr on windows you would do.

command > miopen_output_logs.txt 2>&1

Note that the current situation with ComfyUI is difficult because they disable cudnn by default. There are some nodes that enable it, but I'm not familiar enough with ComfyUI to know what that means. (Just trying to help spread information)

[BradPepersAMD]

MIOpen does output to stderr so that 2>&1 redirect is needed to capture that output and 100% with ComfyUI disabling MIOpen we won't get any logs. I'll check if there is any env var setting to allow turning MIOpen back on or something without it having to be a code change.

[jammm]

Thanks for the catch ! Updated with 2>&1

[stellaraccident]

Will you take gfx1150 or does it need to be gfx1151?

Yes. We're going to optimize all of the commonly used RDNA3/4 kernels and input on any of the architectures is welcome.

[redumarqueze]

Will you take an RDNA 2 GPU (Radeon RX 6600) on KDE Linux?

[stellaraccident]

Will you take an RDNA 2 GPU (Radeon RX 6600) on KDE Linux?

The implementation we are working on right now will not apply automatically to RDNA2 out of the box, but we are interested in not leaving these behind as well. Feedback welcome.

[cjrolo]

Usage from vLLM is relevant?

[stellaraccident]

Usage from vLLM is relevant?

Sure. We've got to optimize all common use cases. This issue was scoped to the end of the yak we are currently approaching with clippers but we need all of it, and it is better based on what people actually are using/need.

[Djip007]

This is not with pytorch, but with llama.cpp. I use llama_bench to cover all pp size from 1 to 2048 (I think it is the best ubatch so not needed higher
I use therock_7.9rc (it is a a bit faster than the new 7.10 "release"

# Use this env:
KV_TYPE=f16
MODEL=Meta-Llama-3.1-8B-Instruct/F16.gguf

export ROCBLAS_USE_HIPBLASLT=1
export GGML_CUDA_ENABLE_UNIFIED_MEMORY=ON
export HIPBLASLT_LOG_MASK=32
export HIPBLASLT_LOG_FILE="Ministral-3-14B-Instruct-2512_BF16-llamacpp_hipblaslt_log.txt"

# and run with:
./bin/llama-bench -ctk ${KV_TYPE} -ctv ${KV_TYPE} -ub 8192 -b 16384  \
  -r 1 -fa 1 -ngl 999 --mmap 0 \
  ${PP} \
  -n 16 \
  -m ${MODEL}

Meta-Llama-3.1-8B-Instruct_FP16-llamacpp_sorted_hipblaslt_log.txt

Meta-Llama-3.1-8B-Instruct_BF16-llamacpp_sorted_hipblaslt_log.txt

and the benchmark result.

Meta-Llama-3.1-8B-Instruct_FP16-llamacpp-bench.log

Meta-Llama-3.1-8B-Instruct_BF16-llamacpp_bench.log

I think of create more if you find it usefull ...

[BellaDoggie]

Hello,

Thank you for addressing these issues! I am getting a failed to upload files error message when trying to add inference and training logs. Any thoughts or suggestion on getting you the log files?

Workload

LLaMA 3.3 70B LoRA fine-tuning and inference for ICD-10 coding

• 343k training samples
• Batch Size 3 Grad Accum = 8
• Training: ~136s/step
• Inference: ~10s/sample

Hardware

AMD Ryzen AI MAX+ 395 (Strix Halo, gfx1151), 96GB unified memory

Known Issues on gfx1151

• SDPA: Forced math backend (flash/mem_efficient broken)
• Unsloth KV cache broken, using HF+PEFT for inference
• Training requires attn_implementation="eager" for causal masking

GEMM Logs Attached

• Training: 10 steps of 70B LoRA fine-tuning
• Inference: 100 samples

ROCm / PyTorch Stack Diagnostic Report

Generated: Wed Dec 17 11:12:50 AM PST 2025

[1] OS + Kernel Info
Distributor ID: Ubuntu
Description: Ubuntu 25.04
Release: 25.04
Codename: plucky
Linux user-box 6.16.12-061612-generic #202510121336 SMP PREEMPT_DYNAMIC Sun Oct 12 16:24:31 UTC 2025 x86_64 x86_64 x86_64 GNU/Linux

[2] ROCm / HIP Info
Name: AMD RYZEN AI MAX+ 395 w/ Radeon 8060S
Marketing Name: AMD RYZEN AI MAX+ 395 w/ Radeon 8060S
Vendor Name: CPU
Name: gfx1151
Marketing Name: AMD Radeon Graphics
Vendor Name: AMD
Name: amdgcn-amd-amdhsa--gfx1151
Name: amdgcn-amd-amdhsa--gfx11-generic
Name: aie2
Marketing Name: AIE-ML
Vendor Name: AMD

7.2.25461-b8eeb4cc53HIP version: 7.2.25461-b8eeb4cc53
AMD clang version 22.0.0git (https://github.com/ROCm/llvm-project.git cf5a4bf00b403b30663fb49a1ac0d2d8add48356+PATCHED:371bd8145f7e4d1d481987d05427c94e17708b91)
Target: x86_64-unknown-linux-gnu
Thread model: posix
InstalledDir: /home/user/miniforge3/envs/therock/lib/python3.12/site-packages/_rocm_sdk_core/lib/llvm/bin

[3] ROCm Packages
rc rocm-opencl 2.0.0.70100-20~24.04 amd64 clr built using CMake

[4] Python Environment Info
/home/user/miniforge3/envs/therock/bin/python
Python 3.12.12
VIRTUAL_ENV:

[5] PyTorch ROCm Info
Torch version: 2.10.0a0+gita5436a5
CUDA available: True
Device: Radeon 8060S Graphics
HIP runtime version: 7.2.25461
BF16 supported: True

[6] bitsandbytes Info
bitsandbytes version: 0.49.0.dev0
bitsandbytes not functional: module 'bitsandbytes' has no attribute 'lib'

[7] Key Python Package Versions
Name: unsloth
Version: 2025.3.19
Name: transformers
Version: 4.57.3
Name: trl
Version: 0.15.2
Name: peft
Version: 0.18.0
Name: datasets
Version: 4.3.0
Name: accelerate
Version: 1.12.0

[8] HIP / HSA Visible Devices
HIP_VISIBLE_DEVICES=0
HSA_VISIBLE_DEVICES=

[9] GPU Compute Test
Success, GPU compute OK. Mean: 0.08141815662384033