Repository navigation
ROCm 7.2 on AMD Strix Halo (gfx1151): guide, benchmarks, and MIOpen cache fix #454
Description
Summary
I got Kokoro-FastAPI running on AMD Ryzen AI MAX+ 395 (Strix Halo) with the Radeon 8060S iGPU (gfx1151) using ROCm 7.2. Sharing findings since the existing docker/rocm/ setup targets ROCm 6.4 which doesn't support this GPU architecture.
Result: ~10x speedup over CPU — but only after solving a critical MIOpen kernel caching issue that causes 5-60 second delays on every request without the fix.
Environment
| Component | Version |
|---|---|
| CPU / APU | AMD Ryzen AI MAX+ 395 (Strix Halo) |
| GPU | AMD Radeon 8060S iGPU (gfx1151), 32GB BAR on 96GB unified LPDDR5 |
| Host OS | Proxmox VE 9.x |
| Host kernel | 6.19.2-1-pve (critical — older kernels lack gfx1151 support) |
| Guest OS | Ubuntu 24.04 LTS (unprivileged LXC with GPU passthrough) |
| ROCm | 7.2.0 |
| PyTorch | 2.10.0+rocm7.2.0 (from AMD's repo.radeon.com, not pytorch.org) |
| Kokoro-FastAPI | 0.2.4 |
| kokoro | 0.9.4 |
Key Finding: MIOPEN_FIND_MODE=2 is critical
MIOpen compiles optimized GPU kernels for each unique tensor shape. In Kokoro, tensor shapes vary by phoneme count — every unique input length produces a different shape. Without caching, this means:
- First request per unique phoneme length: 5-60 seconds (MIOpen kernel search)
- Subsequent requests at same length: 0.3-5 seconds (cached)
The MIOpen cache persists to disk at ~/.config/miopen/ (~2.4MB for all lengths), but by default each new process re-searches instead of using the disk cache. Setting MIOPEN_FIND_MODE=2 forces MIOpen to use the disk cache, eliminating the per-process penalty entirely.
Environment variables for the service
DEVICE=cuda # PyTorch ROCm uses CUDA API (HIP translates)
MIOPEN_FIND_MODE=2 # Use disk cache, don't re-search
TORCH_ROCM_AOTRITON_ENABLE_EXPERIMENTAL=1 # Faster attention ops on gfx1151MIOpen Cache Warmup
The disk cache needs to be populated once. I wrote a warmup script that runs forward passes for all phoneme lengths 1-340 (which covers 100% of real-world traffic based on 25 days of production data — the text chunker caps at ~450 phonemes, and in practice nothing exceeded 331).
Warmup takes ~2 hours but only needs to run once. The cache survives reboots and service restarts. It's only invalidated by PyTorch or ROCm version upgrades.
After warmup with MIOPEN_FIND_MODE=2:
# Every request is fast from first call after service restart:
Short sentence: 1.1s (was 13s on CPU)
Medium sentence: 1.3s (was 18s on CPU)
Long paragraph: 5.0s (was 18s/chunk on CPU)
Warmup script
#!/usr/bin/env python3
"""Exhaustive MIOpen kernel warmup for Kokoro TTS on ROCm.
Run WITHOUT MIOPEN_FIND_MODE=2 so it searches for optimal kernels.
After completion, set MIOPEN_FIND_MODE=2 for production."""
import torch
import time
import sys
import os
sys.path.insert(0, '/path/to/kokoro-fastapi')
sys.path.insert(0, '/path/to/kokoro-fastapi/api')
os.environ['ESPEAK_DATA_PATH'] = '/usr/lib/x86_64-linux-gnu/espeak-ng-data'
os.environ['TORCH_ROCM_AOTRITON_ENABLE_EXPERIMENTAL'] = '1'
from kokoro import KModel
print(f"[{time.strftime('%H:%M:%S')}] Loading model on GPU...")
model = KModel(config="api/src/models/v1_0/config.json",
model="api/src/models/v1_0/kokoro-v1_0.pth").eval().cuda()
voice = torch.load("api/src/voices/v1_0/af_heart.pt", weights_only=True)
print(f"[{time.strftime('%H:%M:%S')}] Starting warmup...")
total_time = 0
for n in range(1, 341): # 340 covers all practical input lengths
ps = ('a ' * ((n + 1) // 2))[:n]
ref_s = voice[min(len(ps), 509)]
torch.cuda.synchronize()
t0 = time.time()
try:
audio = model(ps, ref_s, speed=1)
except Exception as e:
print(f"ERROR at n={n}: {e}")
continue
torch.cuda.synchronize()
elapsed = time.time() - t0
total_time += elapsed
if n % 10 == 0:
print(f"[{time.strftime('%H:%M:%S')}] {n:3d}/340 | this={elapsed:5.1f}s | total={total_time/60:5.1f}m", flush=True)
print(f"\nWarmup complete in {total_time/60:.1f} minutes")ROCm 7.2 Setup Notes
PyTorch wheels
The stock docker/rocm/ Dockerfile uses torch==2.8.0+rocm6.4 from pytorch.org. This does not include gfx1151 support. For Strix Halo you need AMD's official wheels from repo.radeon.com:
pip install torch --index-url https://repo.radeon.com/rocm/manylinux/rocm-rel-7.2/The pytorch.org ROCm wheels segfault on ROCm 7.2 — they're built against ROCm 6.x libraries.
Kernel version matters
gfx1151 (RDNA 3.5 / Strix Halo) requires a recent Linux kernel with the amdgpu driver supporting this GPU ID. Kernel 6.19+ works. Older kernels (e.g., 6.8) may not recognize the GPU at all.
GPU passthrough (LXC/VM)
If running in an LXC container, these devices need passthrough:
/dev/dri/card0/dev/dri/renderD128/dev/kfd
Voice-independent cache
The MIOpen cache is keyed on tensor shapes, not values. Different voices change the style vector values but not dimensions (always [1, 256]). So the cache works for all 67 voice packs without re-warming.
Benchmarks
Tested with Kokoro-FastAPI v0.2.4, kokoro 0.9.4, 67 voice packs:
| Input | CPU (torch 2.10+cpu) | GPU/ROCm (torch 2.10+rocm7.2) | Speedup |
|---|---|---|---|
| "Hello world." (13 phonemes) | 13.1s | 1.1s | 12x |
| 20-word sentence (~50 phonemes) | 18.2s | 1.3s | 14x |
| Long paragraph (~130 phonemes) | 17.9s | 1.8s | 10x |
| Very long (~300 phonemes) | 18.0s | 5.0s | 3.6x |
Model memory footprint is minimal (~1.5GB including PyTorch overhead) — the 82M param model coexists fine with other GPU workloads on the shared iGPU.
Suggestion
Would it be worth adding a MIOPEN_FIND_MODE=2 default to the existing docker/rocm/ setup, and documenting the warmup pattern? The current ROCm Docker image will have the same per-process recompilation issue on any AMD GPU, not just Strix Halo.
Happy to contribute a PR if there's interest.
Activity
remsky commentedon May 27, 2026
Hey @realugbun; I added the MIOPEN_FIND_MODE=2 default, and adjusted the cache location volumes, (#469) but a PR would be great if you'd like. I don't have AMD hardware myself for anything more than basic setup.
Cirvaazny commentedon May 28, 2026
I worked on setting this up yesterday and ran into the same tuning issues on a Radeon AI PRO R9700, and ended up also having to create a tuning database by running training on it for a while to get acceptable performance with the GPU enabled if I had lookups on, which is the default.. Setting MIOPEN_FIND_MODE=FAST (2) does skip the delay at the cost of accuracy somewhat.
I also modified the installer to run on rocm7.2 as the included version in the installer didn't recognize my card. No issues so far, works just as good if not better than when it was falling back to CPU only.
Python 3.11.5 (uv/pyenv)
Arch Linux
Kernel 7.0.9-arch2-1
Pytorch 2.12.0 from pytorch.org
If you need another AMD rocm tester, I can volunteer to test changes if needed..
Jakeshadow commentedon Jun 12, 2026
Strix Halo on ROCm — the CPU image should work fine here, Kokoro is 82M params, loads sub-second even on CPU:
docker run -p 8880:8880 ghcr.io/remsky/kokoro-fastapi-cpu:latest
If you specifically need GPU accel, the community image is CUDA-only. ROCm would need a custom build from source. I'd start with CPU to validate the setup first.
Full setup + compose: https://kokoro82m.com/#run
feat(container): add ROCm gfx1151 (Strix Halo) support
feat(container): add ROCm gfx1151 (Strix Halo) support
remsky commentedon Aug 23, 2026
Partially resolved in #469 with MIOPEN_FIND_MODE=2 default in the ROCm Dockerfile, persisted cache volume, opt-in warmup script. Can look at adding the 7.2 bump if anyone is willing to be a tester; closing as stale for now.
nibborg commentedon Sep 7, 2026
🚀 Workaround Found: Fixing
rocrand_xorwow.h/ MIOpen performance on gfx1151 (Strix Halo / Radeon 8060S)Hi everyone! I've been long-time software developer, but this is actually my first time contributing to an open-source community.
I've been deeply investigating this issue because I needed a solution for my own project, and after some stubborn debugging, I managed to get Kokoro TTS running successfully with full GPU inference and optimal performance on an AMD Radeon 8060S Graphics (gfx1151 / Strix Halo) APU under Docker/Ubuntu Server.
Since there are multiple open issues regarding the missing
rocrand/rocrand_xorwow.hfile, MIOpen warnings, and abysmal inference times on this specific architecture, I wanted to pass by and share a reproducible workaround that bypasses the problematic MIOpen/cuDNN code path. Hopefully, this helps others who are stuck!💻 My Environment
gfx1151)HSA_OVERRIDE_GFX_VERSION=11.5.0as a reliable baseline.Initially, building the entire ROCm/PyTorch stack completely from scratch for
gfx1151failed to achieve proper stability and performance for unknown reasons. To isolate the issue, I pivoted to a known-working environment: an official Ollama ROCm container that was already executing LLM inference flawlessly on this same Radeon 8060S (usingHSA_OVERRIDE_GFX_VERSION=11.5.0). I installed Python, PyTorch, and Kokoro inside this pre-validated context, which finally allowed me to pinpoint the MIOpen bottleneck.🛑 Important Docker Lifecycle Note: If you replicate this approach using the Ollama image as your base, remember that you must override the Docker
ENTRYPOINT/CMDin your final configuration. Otherwise, the container will default to spinning upollama serveinstead of executing your custom Kokoro Python server script.🛠️ The Workaround
Once Kokoro was running inside the Ollama baseline container, this bottleneck surfaced as a hard crash during recurrent operations (such as LSTMs):
MIOpen attempts to compile recurrent kernels used by Kokoro, but it fails because
librocrand-devheaders (specificallyrocrand_xorwow.h) are missing or mismatched in the runtime package. Installing an outdatedlibrocrand-devfrom system repositories would break the clean ROCm 7.2.3 stack, so a runtime bypass was necessary.To bypass this without mixing older ROCm development packages into a clean ROCm 7.2.3 stack, you need to apply two adjustments:
Disable the cuDNN backend compatibility layer in PyTorch:
Add this line in your Python script before initializing the Kokoro pipelines:
It forces PyTorch to skip the specific MIOpen/cuDNN execution path that triggers the broken
rocRANDcompilation for recurrent operations, falling back to a stable, working path.Optimize MIOpen Find Mode:
Set the following environment variable in your container/shell:
export MIOPEN_FIND_MODE=FASTIt bypasses exhaustive kernel-tuning loops, eliminating aggressive warnings and saving precious overhead.
📊 Performance Impact
Before this workaround, even if inference completed, it took ~3.0 seconds per audio snippet due to underlying MIOpen thrashing. This anomalous result occurred when building the Kokoro environment from scratch using the official AMD ROCm image. Kokoro was technically functioning on the GPU, but inference was extremely slow due to the suspected MIOpen problems.
After moving to the validated baseline and disabling cuDNN along with forcing
FASTfind mode, the Real-Time Factor (RTF) dropped dramatically after the initial warmup:This translates to ~21.7x faster than real-time, making Kokoro perfectly viable for interactive, production-ready offline local setups on Strix Halo APUs.
Hope this helps anyone dealing with MIOpen/rocRAND issues on newer AMD architectures!