miopenStatusUnknownError isn't a GPU limitation. It's a missing header.

For nine days I believed my GPU couldn't run batch normalization. I'd tested it, hit a hard runtime error, tried the obvious flags, and moved the workload to CPU with a note in the roadmap explaining the limitation. Voice cloning went from something interactive to something that took about five minutes a sentence.

The GPU was fine. The container was missing libstdc++-dev.

That's the whole bug. It's worth writing up because the error you get tells you almost nothing, the natural conclusion it pushes you toward is wrong, and the wrong conclusion is expensive — it doesn't crash anything, it just quietly makes you architect around a limit that doesn't exist.

The symptom

The machine is an AMD Ryzen AI Max+ 395 — Strix Halo, Radeon 8060S integrated graphics, gfx1151 — running PyTorch 2.10 on ROCm 7.13 from AMD's per-architecture nightly wheels. Any BatchNorm2d forward pass on the GPU raised:

RuntimeError: miopenStatusUnknownError

That's it. No file, no operation, no hint. It failed the same way in eval mode under torch.no_grad(), so it wasn't a training-only path. It failed on a trivial tensor, so it wasn't a memory or shape problem. MIOPEN_FIND_MODE=FAST changed nothing.

Everything about that reads as this chip can't do this op. gfx1151 isn't on AMD's official ROCm support matrix — it works, and the nightly wheels are built for it, but it lives in the unsupported-but-functional zone where you half-expect gaps. So I wrote the limitation down, pinned the voice encoder to CPU, and moved on.

What was actually happening

MIOpen — AMD's deep learning primitives library, the ROCm counterpart to cuDNN — doesn't ship every kernel precompiled. Some are JIT-compiled at runtime, in-process, the first time you hit a given operation and tensor shape. MIOpenBatchNormFwdInferSpatial.cpp is one of them.

That runtime compile is a real C++ compile. It needs the C++ standard library headers to be present inside the container.

My image was ubuntu:24.04, a Python venv, and pip-installed ROCm wheels. Nothing in that chain installs libstdc++-*-dev. So the compile hit #include <type_traits>, found nothing, and failed. MIOpen caught a failed code-object build and PyTorch surfaced it as the only thing it had: an unknown status code.

The actual message was there the whole time. It just wasn't in the exception:

/tmp/comgr-xxxxxx/include/miopen_type_traits.hpp:151:10:
    fatal error: 'type_traits' file not found
MIOpen Error: ...hipoc_program.cpp:299: Code object build failed.
    Source: MIOpenBatchNormFwdInferSpatial.cpp

MIOpen prints that as a warning before raising. In a container log interleaved with model-loading chatter, above a Python traceback that ends in a clean-looking RuntimeError, it scrolls right past. I'd been reading the bottom of the output. The diagnosis was in the middle.

The fix

One package, one layer:

# LOAD-BEARING, not a build tool. Placed AFTER the torch install so adding it
# never invalidates the multi-gigabyte ROCm wheel layer.
RUN apt-get update && apt-get install -y --no-install-recommends libstdc++-13-dev \
    && rm -rf /var/lib/apt/lists/*

Layer order matters more than it looks. Put that line before the pip install and every future edit to it re-downloads several gigabytes of wheels.

Verifying it takes about ten seconds:

docker exec <service> python3 -c "
import torch
x = torch.randn(8, 64, 96, 96, device='cuda')
bn = torch.nn.BatchNorm2d(64).to('cuda').eval()
with torch.no_grad(): y = bn(x)
torch.cuda.synchronize(); print('OK', y.shape)"

Before: miopenStatusUnknownError. After: returns in about 0.21 seconds, including the first-use compile.

One trap while testing this — docker compose build doesn't replace a running container. I spent a few minutes convinced the fix hadn't worked because I was still exec'ing into the old image.

If you can't rebuild the image, there's a one-line bypass:

torch.backends.cudnn.enabled = False   # ROCm maps this onto MIOpen

That forces PyTorch's native batch_norm and skips MIOpen entirely — no JIT compile at all. On the same tensor it ran in 0.020 seconds, an order of magnitude faster than the MIOpen path, because MIOpen's tuned kernels don't earn their compile cost at that size. Useful as a diagnostic too: if disabling it makes the error disappear, your problem is in kernel compilation, not in the GPU or the op.

The second bug hiding behind the first

With headers in place, the first GPU call at an unfamiliar input size looked like a hang — one CPU core pegged at 100%, no GPU activity, no output. I measured over eleven minutes on a 1748×2432 input.

That's not a hang, it's the same JIT mechanism succeeding slowly. Compilation is per shape, and it's CPU-bound. Two things fix it:

1. Persist the kernel cache. MIOpen caches compiled kernels in $HOME/.cache/miopen by default, which in a container means the writable layer — destroyed on every recreate. Mount it on a volume and you pay the compile once, ever.
2. Cap the working resolution. Standardizing input sizes means jobs reuse a small set of shapes instead of compiling a fresh kernel for every image. Capping had a free side effect: it forced even frame dimensions, which yuv420p and libx264 quietly require — odd dimensions kill the encode with no useful error either.

Warm lip-sync of an 8.7-second clip went from 5 minutes 16 seconds on the first uncached run to 1.4 seconds.

What it cost, and what it bought

The voice number is the one that stings. A thirty-fold speedup was sitting behind an apt package for nine days, and the only thing standing in the way was that I'd believed a conclusion I wrote myself.

The part that generalizes

I went looking to see whether I'd found something new. I hadn't, exactly — and the shape of what's already out there is the actually useful finding.

Back in 2023, someone hit the same class of failure with a different header — 'limits' file not found — fixed it with libstdc++-12-dev, and suggested ROCm should declare the dependency so nobody else would trip on it. It still isn't declared. As of this writing there's an open issue reporting 'type_traits' file not found from BatchNorm2D on a different GPU on Windows, sitting in triage with no fix documented.

Then there's the one that changes how I read these errors. There's another open issue reporting miopenStatusUnknownError on batch normalization, on the same gfx1151 chip, from the same wheel source. Same error string, same operation, same architecture — and a completely different root cause. That one is invalid inline assembly operands in MIOpen's reduction kernels. No header would have fixed it.

So miopenStatusUnknownError isn't a diagnosis. It's a category: something failed while producing a kernel. Two people can hit that exact string on that exact chip doing that exact operation and need entirely different fixes. Which means matching on the error text — searching it, comparing it to someone else's thread, concluding you have their problem — is close to worthless. The only thing that discriminates is the compiler output above the exception.

Three habits came out of this, and they're not really about ROCm:

Read upward from the exception. The traceback is where the error surfaced. On anything with a runtime compiler underneath — JIT kernels, shader pipelines, generated SQL, template engines — the actionable message is emitted earlier, by a different layer, in a different format, and usually as a warning nobody's watching.

Treat "this hardware can't" as a hypothesis with an expiration date. It's an unusually comfortable conclusion because it's unfalsifiable from where you're standing and it justifies the workaround you already built. I'd written mine into a roadmap doc, which made it look researched. Nine days later it was just a wrong note that other decisions were stacking on top of.

Test the mechanism, not the symptom. Toggling cudnn.enabled took one line and would have isolated it to kernel compilation on day one. I never ran it, because I'd already stopped asking why and started asking what do I do instead.

The gap between those two questions was five minutes a sentence.

Sources

1. ROCm/MIOpen #2049 — 'limits file not found' — same class of failure, resolved with libstdc++-12-dev
2. ROCm/ROCm #6150 — BatchNorm2D causes miopenStatusUnknownError — 'type_traits' not found on gfx1200, open as of this writing
3. ROCm/TheRock #2488 — MIOpen(HIP) error on gfx1151 PyTorch — identical error string, completely different root cause
4. AMD MIOpen documentation — kernel cache — runtime compilation and the $HOME/.cache/miopen default
5. ROCm/TheRock discussion #655 — per-architecture PyTorch wheels — gfx1151 wheel source

Verified July 28, 2026.