HIPIFY
HIPIFY copied to clipboard
ROCm
[HIPIFY] hipify-clang terminates with constexpr function cannot overload function with same signature
I am trying to convert a CUDA application (https://github.com/mcf-rocks/solanity) with hipify-clang and getting the following error:
/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/cmath:1097:3: error: constexpr function 'fpclassify' without __host__ or __device__ attributes cannot overload __device__ function with same signature. Add a __host__ attribute, or build with -fno-cuda-host-device-constexpr.
I am on arch linux, installed everything from aur, tried with several CUDA versions already.
command:
/opt/rocm/hip/bin/hipify-clang -v --inplace common.cu --cuda-path=/opt/cuda-9.2 -I /opt/rocm/llvm/lib/clang/17.0.0/include/
output:
clang version 17.0.0
Target: x86_64-pc-linux-gnu
Thread model: posix
InstalledDir: /opt/rocm/bin
Found candidate GCC installation: /usr/lib/gcc/x86_64-pc-linux-gnu/11.4.0
Found candidate GCC installation: /usr/lib/gcc/x86_64-pc-linux-gnu/12.3.0
Found candidate GCC installation: /usr/lib/gcc/x86_64-pc-linux-gnu/13.2.1
Found candidate GCC installation: /usr/lib/gcc/x86_64-pc-linux-gnu/7.5.0
Found candidate GCC installation: /usr/lib64/gcc/x86_64-pc-linux-gnu/11.4.0
Found candidate GCC installation: /usr/lib64/gcc/x86_64-pc-linux-gnu/12.3.0
Found candidate GCC installation: /usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1
Found candidate GCC installation: /usr/lib64/gcc/x86_64-pc-linux-gnu/7.5.0
Selected GCC installation: /usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1
Candidate multilib: .;@m64
Candidate multilib: 32;@m32
Selected multilib: .;@m64
Found CUDA installation: /opt/cuda-9.2, version 9.2
Found HIP installation: /opt/rocm, version 5.7.31921
clang Invocation:
"/opt/rocm/bin/clang-tool" "-cc1" "-triple" "x86_64-pc-linux-gnu" "-target-sdk-version=9.2" "-aux-triple" "nvptx64-nvidia-cuda" "-fsyntax-only" "-disable-free" "-clear-ast-before-backend" "-main-file-name" "common.cu-6986b4.hip" "-mrelocation-model" "pic" "-pic-level" "2" "-pic-is-pie" "-mframe-pointer=all" "-fmath-errno" "-ffp-contract=on" "-fno-rounding-math" "-mconstructor-aliases" "-funwind-tables=2" "-target-cpu" "x86-64" "-tune-cpu" "generic" "-debugger-tuning=gdb" "-v" "-fcoverage-compilation-dir=/home/le/src/solanity/src/cuda-ecc-ed25519" "-resource-dir" "/opt/rocm/llvm/lib/clang/17" "-internal-isystem" "/opt/rocm/llvm/lib/clang/17/include/cuda_wrappers" "-include" "__clang_cuda_runtime_wrapper.h" "-I" "/home/le/src/solanity/src/cuda-ecc-ed25519" "-I" "/opt/rocm/llvm/lib/clang/17.0.0/include/" "-I/opt/rocm/hip/include/" "-I." "-internal-isystem" "/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1" "-internal-isystem" "/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/x86_64-pc-linux-gnu" "-internal-isystem" "/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/backward" "-internal-isystem" "/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1" "-internal-isystem" "/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/x86_64-pc-linux-gnu" "-internal-isystem" "/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/backward" "-internal-isystem" "/opt/rocm/llvm/lib/clang/17/include" "-internal-isystem" "/usr/local/include" "-internal-isystem" "/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../x86_64-pc-linux-gnu/include" "-internal-externc-isystem" "/include" "-internal-externc-isystem" "/usr/include" "-internal-isystem" "/opt/rocm/llvm/lib/clang/17/include" "-internal-isystem" "/usr/local/include" "-internal-isystem" "/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../x86_64-pc-linux-gnu/include" "-internal-externc-isystem" "/include" "-internal-externc-isystem" "/usr/include" "-internal-isystem" "/opt/cuda-9.2/include" "-Wno-pragma-once-outside-header" "-std=c++14" "-fdeprecated-macro" "-fdebug-compilation-dir=/home/le/src/solanity/src/cuda-ecc-ed25519" "-ferror-limit" "19" "-fgnuc-version=4.2.1" "-fcxx-exceptions" "-fexceptions" "-fcolor-diagnostics" "-internal-isystem" "/opt/rocm/bin/../include/hipify/cuda_wrappers" "-internal-isystem" "/opt/rocm/bin/../include/hipify" "-internal-isystem" "/opt/rocm/bin/include/cuda_wrappers" "-internal-isystem" "/opt/rocm/bin/include" "-faddrsig" "-D__GCC_HAVE_DWARF2_CFI_ASM=1" "-x" "cuda" "/tmp/common.cu-6986b4.hip"
clang -cc1 version 17.0.0 based upon LLVM 17.0.0git default target x86_64-pc-linux-gnu
ignoring nonexistent directory "/opt/rocm/llvm/lib/clang/17/include/cuda_wrappers"
ignoring nonexistent directory "/opt/rocm/llvm/lib/clang/17/include"
ignoring nonexistent directory "/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../x86_64-pc-linux-gnu/include"
ignoring nonexistent directory "/include"
ignoring nonexistent directory "/opt/rocm/llvm/lib/clang/17/include"
ignoring nonexistent directory "/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../x86_64-pc-linux-gnu/include"
ignoring nonexistent directory "/include"
ignoring nonexistent directory "/opt/rocm/bin/../include/hipify/cuda_wrappers"
ignoring nonexistent directory "/opt/rocm/bin/../include/hipify"
ignoring nonexistent directory "/opt/rocm/bin/include/cuda_wrappers"
ignoring nonexistent directory "/opt/rocm/bin/include"
ignoring duplicate directory "/home/le/src/solanity/src/cuda-ecc-ed25519"
ignoring duplicate directory "/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1"
ignoring duplicate directory "/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/x86_64-pc-linux-gnu"
ignoring duplicate directory "/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/backward"
ignoring duplicate directory "/usr/local/include"
ignoring duplicate directory "/usr/include"
#include "..." search starts here:
#include <...> search starts here:
/home/le/src/solanity/src/cuda-ecc-ed25519
/opt/rocm/llvm/lib/clang/17.0.0/include
/opt/rocm/hip/include
/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1
/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/x86_64-pc-linux-gnu
/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/backward
/usr/local/include
/usr/include
/opt/cuda-9.2/include
End of search list.
In file included from <built-in>:1:
In file included from /opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_runtime_wrapper.h:41:
/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/cmath:1097:3: error: constexpr function 'fpclassify' without __host__ or __device__ attributes cannot overload __device__ function with same signature. Add a __host__ attribute, or build with -fno-cuda-host-device-constexpr.
fpclassify(float __x)
^
/opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_math_forward_declares.h:225:9: note: conflicting __device__ function declared here
using ::fpclassify;
^
In file included from <built-in>:1:
In file included from /opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_runtime_wrapper.h:41:
/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/cmath:1102:3: error: constexpr function 'fpclassify' without __host__ or __device__ attributes cannot overload __device__ function with same signature. Add a __host__ attribute, or build with -fno-cuda-host-device-constexpr.
fpclassify(double __x)
^
/opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_math_forward_declares.h:225:9: note: conflicting __device__ function declared here
using ::fpclassify;
^
In file included from <built-in>:1:
In file included from /opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_runtime_wrapper.h:41:
/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/cmath:1122:3: error: constexpr function 'isfinite' without __host__ or __device__ attributes cannot overload __device__ function with same signature. Add a __host__ attribute, or build with -fno-cuda-host-device-constexpr.
isfinite(float __x)
^
/opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_math_forward_declares.h:229:9: note: conflicting __device__ function declared here
using ::isfinite;
^
In file included from <built-in>:1:
In file included from /opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_runtime_wrapper.h:41:
/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/cmath:1126:3: error: constexpr function 'isfinite' without __host__ or __device__ attributes cannot overload __device__ function with same signature. Add a __host__ attribute, or build with -fno-cuda-host-device-constexpr.
isfinite(double __x)
^
/opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_math_forward_declares.h:229:9: note: conflicting __device__ function declared here
using ::isfinite;
^
In file included from <built-in>:1:
In file included from /opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_runtime_wrapper.h:41:
/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/cmath:1144:3: error: constexpr function 'isinf' without __host__ or __device__ attributes cannot overload __device__ function with same signature. Add a __host__ attribute, or build with -fno-cuda-host-device-constexpr.
isinf(float __x)
^
/opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_math_forward_declares.h:232:9: note: conflicting __device__ function declared here
using ::isinf;
^
In file included from <built-in>:1:
In file included from /opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_runtime_wrapper.h:41:
/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/cmath:1152:3: error: constexpr function 'isinf' without __host__ or __device__ attributes cannot overload __device__ function with same signature. Add a __host__ attribute, or build with -fno-cuda-host-device-constexpr.
isinf(double __x)
^
/opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_math_forward_declares.h:232:9: note: conflicting __device__ function declared here
using ::isinf;
^
In file included from <built-in>:1:
In file included from /opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_runtime_wrapper.h:41:
/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/cmath:1171:3: error: constexpr function 'isnan' without __host__ or __device__ attributes cannot overload __device__ function with same signature. Add a __host__ attribute, or build with -fno-cuda-host-device-constexpr.
isnan(float __x)
^
/opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_math_forward_declares.h:236:9: note: conflicting __device__ function declared here
using ::isnan;
^
In file included from <built-in>:1:
In file included from /opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_runtime_wrapper.h:41:
/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/cmath:1179:3: error: constexpr function 'isnan' without __host__ or __device__ attributes cannot overload __device__ function with same signature. Add a __host__ attribute, or build with -fno-cuda-host-device-constexpr.
isnan(double __x)
^
/opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_math_forward_declares.h:236:9: note: conflicting __device__ function declared here
using ::isnan;
^
In file included from <built-in>:1:
In file included from /opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_runtime_wrapper.h:41:
/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/cmath:1198:3: error: constexpr function 'isnormal' without __host__ or __device__ attributes cannot overload __device__ function with same signature. Add a __host__ attribute, or build with -fno-cuda-host-device-constexpr.
isnormal(float __x)
^
/opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_math_forward_declares.h:237:9: note: conflicting __device__ function declared here
using ::isnormal;
^
In file included from <built-in>:1:
In file included from /opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_runtime_wrapper.h:41:
/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/cmath:1202:3: error: constexpr function 'isnormal' without __host__ or __device__ attributes cannot overload __device__ function with same signature. Add a __host__ attribute, or build with -fno-cuda-host-device-constexpr.
isnormal(double __x)
^
/opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_math_forward_declares.h:237:9: note: conflicting __device__ function declared here
using ::isnormal;
^
In file included from <built-in>:1:
In file included from /opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_runtime_wrapper.h:41:
/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/cmath:1221:3: error: constexpr function 'signbit' without __host__ or __device__ attributes cannot overload __device__ function with same signature. Add a __host__ attribute, or build with -fno-cuda-host-device-constexpr.
signbit(float __x)
^
/opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_math_forward_declares.h:264:9: note: conflicting __device__ function declared here
using ::signbit;
^
In file included from <built-in>:1:
In file included from /opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_runtime_wrapper.h:41:
/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/cmath:1225:3: error: constexpr function 'signbit' without __host__ or __device__ attributes cannot overload __device__ function with same signature. Add a __host__ attribute, or build with -fno-cuda-host-device-constexpr.
signbit(double __x)
^
/opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_math_forward_declares.h:264:9: note: conflicting __device__ function declared here
using ::signbit;
^
In file included from <built-in>:1:
In file included from /opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_runtime_wrapper.h:41:
/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/cmath:1243:3: error: constexpr function 'isgreater' without __host__ or __device__ attributes cannot overload __device__ function with same signature. Add a __host__ attribute, or build with -fno-cuda-host-device-constexpr.
isgreater(float __x, float __y)
^
/opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_math_forward_declares.h:230:9: note: conflicting __device__ function declared here
using ::isgreater;
^
In file included from <built-in>:1:
In file included from /opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_runtime_wrapper.h:41:
/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/cmath:1247:3: error: constexpr function 'isgreater' without __host__ or __device__ attributes cannot overload __device__ function with same signature. Add a __host__ attribute, or build with -fno-cuda-host-device-constexpr.
isgreater(double __x, double __y)
^
/opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_math_forward_declares.h:230:9: note: conflicting __device__ function declared here
using ::isgreater;
^
In file included from <built-in>:1:
In file included from /opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_runtime_wrapper.h:41:
/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/cmath:1269:3: error: constexpr function 'isgreaterequal' without __host__ or __device__ attributes cannot overload __device__ function with same signature. Add a __host__ attribute, or build with -fno-cuda-host-device-constexpr.
isgreaterequal(float __x, float __y)
^
/opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_math_forward_declares.h:231:9: note: conflicting __device__ function declared here
using ::isgreaterequal;
^
In file included from <built-in>:1:
In file included from /opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_runtime_wrapper.h:41:
/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/cmath:1273:3: error: constexpr function 'isgreaterequal' without __host__ or __device__ attributes cannot overload __device__ function with same signature. Add a __host__ attribute, or build with -fno-cuda-host-device-constexpr.
isgreaterequal(double __x, double __y)
^
/opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_math_forward_declares.h:231:9: note: conflicting __device__ function declared here
using ::isgreaterequal;
^
In file included from <built-in>:1:
In file included from /opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_runtime_wrapper.h:41:
/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/cmath:1295:3: error: constexpr function 'isless' without __host__ or __device__ attributes cannot overload __device__ function with same signature. Add a __host__ attribute, or build with -fno-cuda-host-device-constexpr.
isless(float __x, float __y)
^
/opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_math_forward_declares.h:233:9: note: conflicting __device__ function declared here
using ::isless;
^
In file included from <built-in>:1:
In file included from /opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_runtime_wrapper.h:41:
/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/cmath:1299:3: error: constexpr function 'isless' without __host__ or __device__ attributes cannot overload __device__ function with same signature. Add a __host__ attribute, or build with -fno-cuda-host-device-constexpr.
isless(double __x, double __y)
^
/opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_math_forward_declares.h:233:9: note: conflicting __device__ function declared here
using ::isless;
^
In file included from <built-in>:1:
In file included from /opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_runtime_wrapper.h:41:
/usr/lib64/gcc/x86_64-pc-linux-gnu/13.2.1/../../../../include/c++/13.2.1/cmath:1321:3: error: constexpr function 'islessequal' without __host__ or __device__ attributes cannot overload __device__ function with same signature. Add a __host__ attribute, or build with -fno-cuda-host-device-constexpr.
islessequal(float __x, float __y)
^
/opt/rocm/llvm/lib/clang/17.0.0/include/__clang_cuda_math_forward_declares.h:234:9: note: conflicting __device__ function declared here
using ::islessequal;
^
fatal error: too many errors emitted, stopping now [-ferror-limit=]
20 errors generated when compiling for host.
Error while processing /tmp/common.cu-6986b4.hip.
Thanks for your answer. Can't test before January since I am traveling over Christmas 🎄. Will try and report back as soon as I am back in office!
Without the Include parameter, I am getting the following error:
<built-in>:1:10: fatal error: '__clang_cuda_runtime_wrapper.h' file not found
#include "__clang_cuda_runtime_wrapper.h"
The problem is that hipify-clang is awaiting clang header files in the 17 subfolder (in the newest format), whereas the rocm has an older format of this clang's subfolder: 17.0.0:
ignoring nonexistent directory "/opt/rocm/llvm/lib/clang/17/include/cuda_wrappers"
ignoring nonexistent directory "/opt/rocm/llvm/lib/clang/17/include"
What version of the ROCm and HIPIFY packages did you install? Do they have the same version?
As an ugly workaround to make it work in your current configuration, you may copy the contains of the /opt/rocm/llvm/lib/clang/17.0.0/include/ to newly created /opt/rocm/llvm/lib/clang/17/include/. Explicit includes to clang like -I /opt/rocm/llvm/lib/clang/17/include/ are not needed, moreover, they can't help, as instead multiple -internal-isystem includes are used internally by hipify-clang itself. That is why you are getting multiple ignoring nonexistent directory messages in the verbose mode.
I am in arch installing both packages from extra repository. Current package version is (both):
5.7.1-1
If I symlink 17.0.0 to 17 in /opt/rocm/llvm/lib/clang/, I seem to be able to convert a file with the following command:
/opt/rocm/hip/bin/hipify-clang -v --inplace common.cu --cuda-path=/opt/cuda -I /opt/rocm/include/
Moving on to a more complex file, however, seems to greate more issues:
/tmp/aes_cbc.cu-6884ee.hip:199:16: error: redefinition of 'l_te' with a different type: 'const u32 *' (aka 'const unsigned int *') vs 'u32[256]' (aka 'unsigned int[256]')
const u32* l_te = g_Te0;
^
/tmp/aes_cbc.cu-6884ee.hip:194:20: note: previous definition is here
__shared__ u32 l_te[256];
^
7 warnings and 1 error generated when compiling for host.
Error while processing /tmp/aes_cbc.cu-6884ee.hip.
I have to admit, that I have no / little experience with cuda and GPU programming at all, so this might take more time to investigate (If it even makes sense what I am doing) :/
Thanks for your help anyways! :)
Moving on to a more complex file, however, seems to greate more issues:
/tmp/aes_cbc.cu-6884ee.hip:199:16: error: redefinition of 'l_te' with a different type: 'const u32 *' (aka 'const unsigned int *') vs 'u32[256]' (aka 'unsigned int[256]') const u32* l_te = g_Te0; ^ /tmp/aes_cbc.cu-6884ee.hip:194:20: note: previous definition is here __shared__ u32 l_te[256]; ^ 7 warnings and 1 error generated when compiling for host. Error while processing /tmp/aes_cbc.cu-6884ee.hip.I have to admit, that I have no / little experience with
cudaand GPU programming at all, so this might take more time to investigate (If it even makes sense what I am doing) :/Thanks for your help anyways! :)
Well, the compiler says that you have a redefinition of u32, and it looks like it is redefined. Have you tried to compile your CUDA source by CUDA's nvcc? Is the source correct?
Sorry for never circling back on this. Just did not find the time to continue investigating... :( Thanks for your help, though!