| Issue |
207801
|
| Summary |
ROCm 7.14 amdclang++ (LLVM 23) miscompiles hand-vectorized kernels at -O3, causing segfaults
|
| Labels |
new issue
|
| Assignees |
|
| Reporter |
kiritigowda
|
## Summary
Building the `openvx` library with **ROCm 7.14** (`amdclang++` / AMD clang 23, based on LLVM 23) at `-O3` miscompiles the hand-vectorized CPU kernels in `amd_openvx/openvx/ago/ago_haf_cpu_*.cpp`, producing out-of-bounds memory accesses and segfaults at runtime. This was not reproducible with ROCm 7.13.
The first kernel to fault is `HafCpu_ChannelCopy_U8_U8`, reached via `vxCopyImagePatch` -> `vxCommitImagePatch`. Several CPU ctest cases crash as a result:
- `openvx_data_objects_api`
- `openvx_graph_import_api`
- `openvx_vxu_api`
- `openvx_coverage_boost`
- `openvx_vision_coverage`
## Diagnosis
The `ago_haf_cpu_*.cpp` kernels are already explicitly vectorized with AVX2/SSE intrinsics. LLVM 23's **loop auto-vectorizer** re-vectorizes these intrinsic loops and generates a broken induction variable. In `HafCpu_ChannelCopy_U8_U8`, called with valid arguments (a 64x48 patch copied into a 640x480 U8 image), the fault occurs at:
```
=> vmovups (%r8,%r9,1),%ymm0
```
with the index register `r9 = 0xf74032d0` (~4 GB) even though `dstWidth` is only 64 — i.e. the loop index ran far past the buffer.
Confirmed via bisection of compiler flags:
| Build | Result |
|-------|--------|
| `-O0` / `-O1` / `-O2` | OK |
| `-O3` | **Segfault** |
| `-O3 -fno-vectorize` | OK |
| `-O3 -fno-slp-vectorize` | Segfault |
| `-O3 -fno-unroll-loops` | Segfault |
| `-O3 -mno-avx2 -msse4.2` | Segfault |
Only disabling the loop auto-vectorizer (`-fno-vectorize`) or dropping below `-O3` avoids the miscompilation, which points at the loop vectorizer re-processing already-vectorized intrinsic code.
## Minimal reproducer
```cpp
// Build (crashes): amdclang++ -std=c++17 -mavx2 -mfma -mbmi2 -O3 channelcopy_repro.cpp -o repro && ./repro
// Build (works): add -fno-vectorize, or use -O2.
#include <immintrin.h>
#include <cstdio>
#include <cstdlib>
__attribute__((noinline))
int HafCpu_ChannelCopy_U8_U8(unsigned int dstWidth, unsigned int dstHeight,
unsigned char *pDstImage, unsigned int dstImageStrideInBytes,
unsigned char *pSrcImage, unsigned int srcImageStrideInBytes)
{
for (unsigned int height = 0; height < dstHeight; height++) {
unsigned char *pLocalSrc = pSrcImage;
unsigned char *pLocalDst = pDstImage;
unsigned int width = 0;
for (; width + 128 <= dstWidth; width += 128) {
__m256i r0 = _mm256_loadu_si256((const __m256i *)(pLocalSrc + width));
__m256i r1 = _mm256_loadu_si256((const __m256i *)(pLocalSrc + width + 32));
__m256i r2 = _mm256_loadu_si256((const __m256i *)(pLocalSrc + width + 64));
__m256i r3 = _mm256_loadu_si256((const __m256i *)(pLocalSrc + width + 96));
_mm256_storeu_si256((__m256i *)(pLocalDst + width), r0);
_mm256_storeu_si256((__m256i *)(pLocalDst + width + 32), r1);
_mm256_storeu_si256((__m256i *)(pLocalDst + width + 64), r2);
_mm256_storeu_si256((__m256i *)(pLocalDst + width + 96), r3);
}
for (; width + 32 <= dstWidth; width += 32) {
__m256i r0 = _mm256_loadu_si256((const __m256i *)(pLocalSrc + width));
_mm256_storeu_si256((__m256i *)(pLocalDst + width), r0);
}
for (; width < dstWidth; width++)
pLocalDst[width] = pLocalSrc[width];
pSrcImage += srcImageStrideInBytes;
pDstImage += dstImageStrideInBytes;
}
return 0;
}
int main() {
// Mirrors vxCopyImagePatch writing a 64x48 patch into a 640x480 U8 image.
const unsigned int W = 64, H = 48, dstStride = 640, srcStride = 64;
unsigned char *dst = (unsigned char *)calloc((size_t)dstStride * 480, 1);
unsigned char *src = "" char *)calloc((size_t)srcStride * H, 1);
for (unsigned i = 0; i < srcStride * H; i++) src[i] = (unsigned char)(i % 256);
HafCpu_ChannelCopy_U8_U8(W, H, dst, dstStride, src, srcStride);
int mismatches = 0;
for (unsigned y = 0; y < H; y++)
for (unsigned x = 0; x < W; x++)
if (dst[y * dstStride + x] != src[y * srcStride + x]) mismatches++;
printf("done, mismatches=%d\n", mismatches);
return 0;
}
```
Output:
```
$ amdclang++ -std=c++17 -mavx2 -mfma -mbmi2 -O3 channelcopy_repro.cpp -o repro && ./repro
Segmentation fault (core dumped) # exit 139
$ amdclang++ -std=c++17 -mavx2 -mfma -mbmi2 -O3 -fno-vectorize channelcopy_repro.cpp -o repro && ./repro
done, mismatches=0
```
## Suggested fix
Add `-fno-vectorize` to the `openvx` target compile options in `amd_openvx/openvx/CMakeLists.txt`. The kernels are already hand-vectorized, so disabling the auto loop-vectorizer for this library has no intended performance cost while avoiding the miscompilation. With this flag the full CPU ctest suite (25/25) passes.
## Environment
- ROCm: `amdrocm-core-sdk7.14-gfx1100` 7.14.0~20260618-27728528519
- Compiler: AMD clang version 23.0.0git (ROCm/llvm-project 2abe93d58c...+PATCHED)
- OS: Ubuntu 24.04.4 LTS, x86_64 (CPU with AVX2 + AVX512)
- MIVisionX: `develop` (4.0.0), CPU/HOST kernels (`USE_AVX=1`)
- Not reproducible on ROCm 7.13.
## Original issue
https://github.com/ROCm/MIVisionX/issues/1707
![]()
_______________________________________________
llvm-bugs mailing list
[email protected]
https://lists.llvm.org/cgi-bin/mailman/listinfo/llvm-bugs
