composable_kernel icon indicating copy to clipboard operation
composable_kernel copied to clipboard

[Issue]: error: invalid operand for instruction while building composable kernel

Open lshqqytiger opened this issue 10 months ago • 3 comments

Problem Description

I'm getting invalid operand for instruction error while compiling inline assembly v_cmpx_le_u32 and buffer_load_dword.

I tried with -DDL_KERNELS=On and Off, but nothing changed.

Am I doing something wrong?

Environment ROCm 6.2.41512-db3292736 MSVC 19.43.34808 Windows SDK 10.0.26100.0

CMake

cmake .. -DCMAKE_BUILD_TYPE=Release -DCMAKE_C_COMPILER="E:/AMD/ROCm/6.2/bin/clang.exe" -DCMAKE_CXX_COMPILER="E:/AMD/ROCm/6.2/bin/clang++.exe" -DCMAKE_PREFIX_PATH="E:/AMD/ROCm/6.2" -G "Ninja" -DCMAKE_MAKE_PROGRAM="ninja.exe" -DHIP_PLATFORM="amd" -DBUILD_TESTING=Off -DCMAKE_HIP_COMPILER_ROCM_ROOT="E:/AMD/ROCm/6.2" -DGPU_TARGETS="gfx908;gfx90a;gfx1030;gfx1100;gfx1101;gfx1102" -DDTYPES="fp64;fp32;fp16;bf16;fp8" -DDL_KERNELS=Off -DDPP_KERNELS=On

Full log including gfx1030 LOG_1030.txt

Full log after excluding gfx1030 LOG.txt

Operating System

Windows 11 (10.0.26100)

CPU

Intel(R) Core(TM) i9-14900K

GPU

AMD Radeon RX 7900 XTX

Other

No response

ROCm Version

ROCm 6.0.0

ROCm Component

Composable Kernel

Steps to Reproduce

No response

(Optional for Linux users) Output of /opt/rocm/bin/rocminfo --support

No response

Additional Information

No response

lshqqytiger avatar Feb 16 '25 04:02 lshqqytiger

Hi @lshqqytiger. Internal ticket has been created to assist with your issue. Thanks!

ppanchad-amd avatar Feb 18 '25 20:02 ppanchad-amd

@lshqqytiger according to the documentation, for gfx1030, you shall use -D GPU_ARCHS : "NOTE: If you try setting GPU_TARGETS to a list of architectures, the build will only work if the architectures are similar, e.g., gfx908;gfx90a, or gfx1100;gfx1101;gfx11012. Otherwise, if you want to build the library for a list of different architectures, you should use the GPU_ARCHS build argument, for example GPU_ARCHS=gfx908;gfx1030;gfx1100;gfx942."

adityas-amd avatar Feb 25 '25 21:02 adityas-amd

Thank you for reply. As you suggest, I replaced -DGPU_TARGETS with -DGPU_ARCHS. However, I still have similar errors. Should I try again with -DGPU_TARGETS and similar architectures? LOG.txt

lshqqytiger avatar Feb 26 '25 22:02 lshqqytiger

This works, thanks. -DGPU_TARGETS="gfx908;gfx1100"

lshqqytiger avatar Apr 24 '25 11:04 lshqqytiger