search

ROCm / HIP-Examples

Examples for HIP
194 stars 86 forks source link

Compilation failure with nvcc backend #21

Open lahwaacz opened 3 years ago

lahwaacz commented 3 years ago

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
datafl4sh commented 3 years ago

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.

datafl4sh commented 3 years ago

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.