HIP-Examples
HIP-Examples copied to clipboard
Published
20 hours ago •
ROCm-Developer-Tools
ROCm-Developer-Tools
Compilation failure with nvcc backend
I'm getting these errors when I try to compile an example with the HIP nvcc backend. What is the problem?
$ cd add4
$ make
/opt/rocm/hip/bin/hipcc -std=c++11 -O3 -c hip-stream.cpp -o hip-stream.o
In file included from /opt/rocm/hip/include/hip/hip_runtime_api.h:368,
from /opt/rocm/hip/include/hip/nvcc_detail/hip_runtime.h:28,
from /opt/rocm/hip/include/hip/hip_runtime.h:58,
from hip-stream.cpp:1:
/opt/rocm/hip/include/hip/nvcc_detail/hip_runtime_api.h: In function ‘hipError_t hipPointerGetAttributes(hipPointerAttribute_t*, const void*)’:
/opt/rocm/hip/include/hip/nvcc_detail/hip_runtime_api.h:1355:21: error: ‘struct cudaPointerAttributes’ has no member named ‘memoryType’
1355 | switch (cPA.memoryType) {
| ^~~~~~~~~~
hip-stream.cpp: In function ‘void copy_looper(const T*, T*, int)’:
hip-stream.cpp:74:19: error: ‘hipBlockIdx_x’ was not declared in this scope
74 | int offset = (hipBlockIdx_x * hipBlockDim_x + hipThreadIdx_x)*CLUMP_SIZE;
| ^~~~~~~~~~~~~
hip-stream.cpp:74:35: error: ‘hipBlockDim_x’ was not declared in this scope
74 | int offset = (hipBlockIdx_x * hipBlockDim_x + hipThreadIdx_x)*CLUMP_SIZE;
| ^~~~~~~~~~~~~
hip-stream.cpp:74:51: error: ‘hipThreadIdx_x’ was not declared in this scope
74 | int offset = (hipBlockIdx_x * hipBlockDim_x + hipThreadIdx_x)*CLUMP_SIZE;
| ^~~~~~~~~~~~~~
hip-stream.cpp:75:34: error: ‘hipGridDim_x’ was not declared in this scope
75 | int stride = hipBlockDim_x * hipGridDim_x * CLUMP_SIZE;
| ^~~~~~~~~~~~
hip-stream.cpp: In function ‘void mul_looper(T*, const T*, int)’:
hip-stream.cpp:86:18: error: ‘hipBlockIdx_x’ was not declared in this scope
86 | int offset = hipBlockIdx_x * hipBlockDim_x + hipThreadIdx_x;
| ^~~~~~~~~~~~~
hip-stream.cpp:86:34: error: ‘hipBlockDim_x’ was not declared in this scope
86 | int offset = hipBlockIdx_x * hipBlockDim_x + hipThreadIdx_x;
| ^~~~~~~~~~~~~
hip-stream.cpp:86:50: error: ‘hipThreadIdx_x’ was not declared in this scope
86 | int offset = hipBlockIdx_x * hipBlockDim_x + hipThreadIdx_x;
| ^~~~~~~~~~~~~~
hip-stream.cpp:87:34: error: ‘hipGridDim_x’ was not declared in this scope
87 | int stride = hipBlockDim_x * hipGridDim_x;
| ^~~~~~~~~~~~
hip-stream.cpp: In function ‘void add_looper(const T*, const T*, const T*, const T*, T*, int)’:
hip-stream.cpp:99:18: error: ‘hipBlockIdx_x’ was not declared in this scope
99 | int offset = hipBlockIdx_x * hipBlockDim_x + hipThreadIdx_x;
| ^~~~~~~~~~~~~
hip-stream.cpp:99:34: error: ‘hipBlockDim_x’ was not declared in this scope
99 | int offset = hipBlockIdx_x * hipBlockDim_x + hipThreadIdx_x;
| ^~~~~~~~~~~~~
hip-stream.cpp:99:50: error: ‘hipThreadIdx_x’ was not declared in this scope
99 | int offset = hipBlockIdx_x * hipBlockDim_x + hipThreadIdx_x;
| ^~~~~~~~~~~~~~
hip-stream.cpp:100:34: error: ‘hipGridDim_x’ was not declared in this scope
100 | int stride = hipBlockDim_x * hipGridDim_x;
| ^~~~~~~~~~~~
hip-stream.cpp: In function ‘void triad_looper(T*, const T*, const T*, int)’:
hip-stream.cpp:111:18: error: ‘hipBlockIdx_x’ was not declared in this scope
111 | int offset = hipBlockIdx_x * hipBlockDim_x + hipThreadIdx_x;
| ^~~~~~~~~~~~~
hip-stream.cpp:111:34: error: ‘hipBlockDim_x’ was not declared in this scope
111 | int offset = hipBlockIdx_x * hipBlockDim_x + hipThreadIdx_x;
| ^~~~~~~~~~~~~
hip-stream.cpp:111:50: error: ‘hipThreadIdx_x’ was not declared in this scope
111 | int offset = hipBlockIdx_x * hipBlockDim_x + hipThreadIdx_x;
| ^~~~~~~~~~~~~~
hip-stream.cpp:112:34: error: ‘hipGridDim_x’ was not declared in this scope
112 | int stride = hipBlockDim_x * hipGridDim_x;
| ^~~~~~~~~~~~
hip-stream.cpp: In function ‘void copy(const T*, T*)’:
hip-stream.cpp:127:19: error: ‘hipBlockDim_x’ was not declared in this scope
127 | const int i = hipBlockDim_x * hipBlockIdx_x + hipThreadIdx_x;
| ^~~~~~~~~~~~~
hip-stream.cpp:127:35: error: ‘hipBlockIdx_x’ was not declared in this scope
127 | const int i = hipBlockDim_x * hipBlockIdx_x + hipThreadIdx_x;
| ^~~~~~~~~~~~~
hip-stream.cpp:127:51: error: ‘hipThreadIdx_x’ was not declared in this scope
127 | const int i = hipBlockDim_x * hipBlockIdx_x + hipThreadIdx_x;
| ^~~~~~~~~~~~~~
hip-stream.cpp: In function ‘void mul(T*, const T*)’:
hip-stream.cpp:137:19: error: ‘hipBlockDim_x’ was not declared in this scope
137 | const int i = hipBlockDim_x * hipBlockIdx_x + hipThreadIdx_x;
| ^~~~~~~~~~~~~
hip-stream.cpp:137:35: error: ‘hipBlockIdx_x’ was not declared in this scope
137 | const int i = hipBlockDim_x * hipBlockIdx_x + hipThreadIdx_x;
| ^~~~~~~~~~~~~
hip-stream.cpp:137:51: error: ‘hipThreadIdx_x’ was not declared in this scope
137 | const int i = hipBlockDim_x * hipBlockIdx_x + hipThreadIdx_x;
| ^~~~~~~~~~~~~~
hip-stream.cpp: In function ‘void add(const T*, const T*, const T*, const T*, T*)’:
hip-stream.cpp:145:19: error: ‘hipBlockDim_x’ was not declared in this scope
145 | const int i = hipBlockDim_x * hipBlockIdx_x + hipThreadIdx_x;
| ^~~~~~~~~~~~~
hip-stream.cpp:145:35: error: ‘hipBlockIdx_x’ was not declared in this scope
145 | const int i = hipBlockDim_x * hipBlockIdx_x + hipThreadIdx_x;
| ^~~~~~~~~~~~~
hip-stream.cpp:145:51: error: ‘hipThreadIdx_x’ was not declared in this scope
145 | const int i = hipBlockDim_x * hipBlockIdx_x + hipThreadIdx_x;
| ^~~~~~~~~~~~~~
hip-stream.cpp: In function ‘void triad(T*, const T*, const T*)’:
hip-stream.cpp:154:19: error: ‘hipBlockDim_x’ was not declared in this scope
154 | const int i = hipBlockDim_x * hipBlockIdx_x + hipThreadIdx_x;
| ^~~~~~~~~~~~~
hip-stream.cpp:154:35: error: ‘hipBlockIdx_x’ was not declared in this scope
154 | const int i = hipBlockDim_x * hipBlockIdx_x + hipThreadIdx_x;
| ^~~~~~~~~~~~~
hip-stream.cpp:154:51: error: ‘hipThreadIdx_x’ was not declared in this scope
154 | const int i = hipBlockDim_x * hipBlockIdx_x + hipThreadIdx_x;
| ^~~~~~~~~~~~~~
In file included from /opt/rocm/hip/include/hip/hip_runtime.h:58,
from hip-stream.cpp:1:
hip-stream.cpp: In function ‘int main(int, char**)’:
hip-stream.cpp:340:17: error: expected primary-expression before ‘<’ token
340 | hipLaunchKernelGGL((copy_looper<float,1>), dim3(gridSize), dim3(groupSize), 0, 0, (float*)d_a, (float*)d_c, ARRAY_SIZE);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:340:17: error: expected primary-expression before ‘>’ token
340 | hipLaunchKernelGGL((copy_looper<float,1>), dim3(gridSize), dim3(groupSize), 0, 0, (float*)d_a, (float*)d_c, ARRAY_SIZE);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:342:17: error: expected primary-expression before ‘<’ token
342 | hipLaunchKernelGGL((copy_looper<double,1>), dim3(gridSize), dim3(groupSize), 0, 0, (double*)d_a, (double*)d_c, ARRAY_SIZE);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:342:17: error: expected primary-expression before ‘>’ token
342 | hipLaunchKernelGGL((copy_looper<double,1>), dim3(gridSize), dim3(groupSize), 0, 0, (double*)d_a, (double*)d_c, ARRAY_SIZE);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:345:17: error: expected primary-expression before ‘<’ token
345 | hipLaunchKernelGGL(copy<float>, dim3(ARRAY_SIZE/groupSize), dim3(groupSize), 0, 0, (float*)d_a, (float*)d_c);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:345:17: error: expected primary-expression before ‘>’ token
345 | hipLaunchKernelGGL(copy<float>, dim3(ARRAY_SIZE/groupSize), dim3(groupSize), 0, 0, (float*)d_a, (float*)d_c);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:347:17: error: expected primary-expression before ‘<’ token
347 | hipLaunchKernelGGL(copy<double>, dim3(ARRAY_SIZE/groupSize), dim3(groupSize), 0, 0, (double*)d_a, (double*)d_c);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:347:17: error: expected primary-expression before ‘>’ token
347 | hipLaunchKernelGGL(copy<double>, dim3(ARRAY_SIZE/groupSize), dim3(groupSize), 0, 0, (double*)d_a, (double*)d_c);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:359:17: error: expected primary-expression before ‘<’ token
359 | hipLaunchKernelGGL(mul_looper<float>, dim3(gridSize), dim3(groupSize), 0, 0, (float*)d_b, (float*)d_c, ARRAY_SIZE);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:359:17: error: expected primary-expression before ‘>’ token
359 | hipLaunchKernelGGL(mul_looper<float>, dim3(gridSize), dim3(groupSize), 0, 0, (float*)d_b, (float*)d_c, ARRAY_SIZE);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:361:17: error: expected primary-expression before ‘<’ token
361 | hipLaunchKernelGGL(mul_looper<double>, dim3(gridSize), dim3(groupSize), 0, 0, (double*)d_b, (double*)d_c, ARRAY_SIZE);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:361:17: error: expected primary-expression before ‘>’ token
361 | hipLaunchKernelGGL(mul_looper<double>, dim3(gridSize), dim3(groupSize), 0, 0, (double*)d_b, (double*)d_c, ARRAY_SIZE);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:364:17: error: expected primary-expression before ‘<’ token
364 | hipLaunchKernelGGL(mul<float>, dim3(ARRAY_SIZE/groupSize), dim3(groupSize), 0, 0, (float*)d_b, (float*)d_c);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:364:17: error: expected primary-expression before ‘>’ token
364 | hipLaunchKernelGGL(mul<float>, dim3(ARRAY_SIZE/groupSize), dim3(groupSize), 0, 0, (float*)d_b, (float*)d_c);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:366:17: error: expected primary-expression before ‘<’ token
366 | hipLaunchKernelGGL(mul<double>, dim3(ARRAY_SIZE/groupSize), dim3(groupSize), 0, 0, (double*)d_b, (double*)d_c);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:366:17: error: expected primary-expression before ‘>’ token
366 | hipLaunchKernelGGL(mul<double>, dim3(ARRAY_SIZE/groupSize), dim3(groupSize), 0, 0, (double*)d_b, (double*)d_c);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:378:17: error: expected primary-expression before ‘<’ token
378 | hipLaunchKernelGGL(add_looper<float>, dim3(gridSize), dim3(groupSize), 0, 0, (float*)d_a, (float*)d_b,(float*)d_d, (float*)d_e, (float*)d_c, ARRAY_SIZE);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:378:17: error: expected primary-expression before ‘>’ token
378 | hipLaunchKernelGGL(add_looper<float>, dim3(gridSize), dim3(groupSize), 0, 0, (float*)d_a, (float*)d_b,(float*)d_d, (float*)d_e, (float*)d_c, ARRAY_SIZE);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:380:17: error: expected primary-expression before ‘<’ token
380 | hipLaunchKernelGGL(add_looper<double>, dim3(gridSize), dim3(groupSize), 0, 0, (double*)d_a, (double*)d_b, (double*)d_d, (double*)d_e,(double*)d_c, ARRAY_SIZE);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:380:17: error: expected primary-expression before ‘>’ token
380 | hipLaunchKernelGGL(add_looper<double>, dim3(gridSize), dim3(groupSize), 0, 0, (double*)d_a, (double*)d_b, (double*)d_d, (double*)d_e,(double*)d_c, ARRAY_SIZE);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:383:17: error: expected primary-expression before ‘<’ token
383 | hipLaunchKernelGGL(add<float>, dim3(ARRAY_SIZE/groupSize), dim3(groupSize), 0, 0, (float*)d_a, (float*)d_b, (float*)d_d,(float*)d_e,(float*)d_c);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:383:17: error: expected primary-expression before ‘>’ token
383 | hipLaunchKernelGGL(add<float>, dim3(ARRAY_SIZE/groupSize), dim3(groupSize), 0, 0, (float*)d_a, (float*)d_b, (float*)d_d,(float*)d_e,(float*)d_c);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:385:17: error: expected primary-expression before ‘<’ token
385 | hipLaunchKernelGGL(add<double>, dim3(ARRAY_SIZE/groupSize), dim3(groupSize), 0, 0, (double*)d_a, (double*)d_b, (double*)d_d,(double*)d_e,(double*)d_c);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:385:17: error: expected primary-expression before ‘>’ token
385 | hipLaunchKernelGGL(add<double>, dim3(ARRAY_SIZE/groupSize), dim3(groupSize), 0, 0, (double*)d_a, (double*)d_b, (double*)d_d,(double*)d_e,(double*)d_c);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:397:17: error: expected primary-expression before ‘<’ token
397 | hipLaunchKernelGGL(triad_looper<float>, dim3(gridSize), dim3(groupSize), 0, 0, (float*)d_a, (float*)d_b, (float*)d_c, ARRAY_SIZE);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:397:17: error: expected primary-expression before ‘>’ token
397 | hipLaunchKernelGGL(triad_looper<float>, dim3(gridSize), dim3(groupSize), 0, 0, (float*)d_a, (float*)d_b, (float*)d_c, ARRAY_SIZE);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:399:17: error: expected primary-expression before ‘<’ token
399 | hipLaunchKernelGGL(triad_looper<double>, dim3(gridSize), dim3(groupSize), 0, 0, (double*)d_a, (double*)d_b, (double*)d_c, ARRAY_SIZE);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:399:17: error: expected primary-expression before ‘>’ token
399 | hipLaunchKernelGGL(triad_looper<double>, dim3(gridSize), dim3(groupSize), 0, 0, (double*)d_a, (double*)d_b, (double*)d_c, ARRAY_SIZE);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:402:17: error: expected primary-expression before ‘<’ token
402 | hipLaunchKernelGGL(triad<float>, dim3(ARRAY_SIZE/groupSize), dim3(groupSize), 0, 0, (float*)d_a, (float*)d_b, (float*)d_c);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:402:17: error: expected primary-expression before ‘>’ token
402 | hipLaunchKernelGGL(triad<float>, dim3(ARRAY_SIZE/groupSize), dim3(groupSize), 0, 0, (float*)d_a, (float*)d_b, (float*)d_c);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:404:17: error: expected primary-expression before ‘<’ token
404 | hipLaunchKernelGGL(triad<double>, dim3(ARRAY_SIZE/groupSize), dim3(groupSize), 0, 0, (double*)d_a, (double*)d_b, (double*)d_c);
| ^~~~~~~~~~~~~~~~~~
hip-stream.cpp:404:17: error: expected primary-expression before ‘>’ token
404 | hipLaunchKernelGGL(triad<double>, dim3(ARRAY_SIZE/groupSize), dim3(groupSize), 0, 0, (double*)d_a, (double*)d_b, (double*)d_c);
| ^~~~~~~~~~~~~~~~~~
make: *** [Makefile:14: hip-stream.o] Error 1
My hipconfig output is:
HIP version : 3.7.
== hipconfig
HIP_PATH : /opt/rocm/hip
ROCM_PATH : /opt/rocm
HIP_COMPILER : clang
HIP_PLATFORM : nvcc
HIP_RUNTIME : CUDA
CPP_CONFIG : -D__HIP_PLATFORM_NVCC__= -I/opt/rocm/hip/include -I/opt/cuda/include
== nvcc
nvcc: NVIDIA (R) Cuda compiler driver
Copyright (c) 2005-2020 NVIDIA Corporation
Built on Wed_Jul_22_19:09:09_PDT_2020
Cuda compilation tools, release 11.0, V11.0.221
Build cuda_11.0_bu.TC445_37.28845127_0
=== Environment Variables
PATH=[censored]
CUDA_CACHE_PATH=[censored]
CUDA_PATH=/opt/cuda
LD_LIBRARY_PATH=[censored]
== Linux Kernel
Hostname : [censored]
Linux [censored] 5.7.10-arch1-1 #1 SMP PREEMPT Wed, 22 Jul 2020 19:57:42 +0000 x86_64 GNU/Linux
Exactly the same thing here...fresh installation from official repositories (3.9.20412-6d111f85) and the compiler complains that all the hipBlockDim_*, hipBlockIdx_*, hipThreadIdx_* variables are not declared.
Exactly the same thing here...fresh installation from official repositories (3.9.20412-6d111f85) and the compiler complains that all the
hipBlockDim_*,hipBlockIdx_*,hipThreadIdx_*variables are not declared.
OK, solved. See https://github.com/ROCm-Developer-Tools/HIP/issues/2163.