使用nvptx-none与gcc链接时发生OpenMP卸载错误:未解析符号_fputwc_r

uurv41yg  于 2023-04-12  发布在  其他
关注(0)|答案(1)|浏览(151)

我正在尝试使用OpenMP卸载为Nvidia GPU编译一个简单的测试问题。我使用gcc和nvptx-none目标。我已经用spack安装了gcc+nvptx包(或者我自己用nvptx-tools编译了gcc-13,结果是一样的)。在链接过程中,我得到错误:

unresolved symbol _fputwc_r
collect2: error: ld returned 1 exit status
mkoffload: fatal error: x86_64-pc-linux-gnu-accel-nvptx-none-gcc returned 1 exit status
compilation terminated.
lto-wrapper: fatal error: /path/to/spack/opt/spack/linux-centos8-x86_64_v3/gcc-13.0.0/gcc-12.2.0-6olbpwbs53cquwnpsvrmuxprmaofwjtk/libexec/gcc/x86_64-pc-linux-gnu/12.2.0//accel/nvptx-none/mkoffload returned 1 exit status
compilation terminated.
/usr/bin/ld: error: lto-wrapper failed

使用-fno-stack-protector编译,如推荐的herehere,并不能缓解这个问题。-fno-lto可以,但卸载不起作用。不同的优化标志没有区别。
使用的ld看起来是系统安装。spack安装在spack/linux-centos8-x86_64_v3/gcc-13.0.0/gcc-12.2.0-6olbpwbs53cquwnpsvrmuxprmaofwjtk/nvptx-none中提供了另一个ld,但spack通常不会将其添加到PATH中。我猜这是有原因的,因为包含它会导致

as: unrecognized option '--64'
nvptx-as: missing .version directive at start of file '/tmp/cc9YfveM.s'``

这是链接器的问题,还是其他问题?问题只发生在实际包含并行for循环时,只是设置#pragma omp target不会。设备实际上被识别,并且根据OpenMP在设备上运行此杂注中的代码,只要不存在并行区域,就会产生上述错误。
其他信息:系统是Rocky Linux release 8.7 (Green Obsidian)我执行的测试程序是基于OpenMP测试程序。它的完整代码是:

#include <stdio.h>
#include <stdlib.h>
#include <omp.h>
void saxpy(float a, float* x, float* y, int sz) {
#pragma omp target teams distribute parallel for simd \
   num_teams(3) map(to:x[0:sz]) map(tofrom:y[0:sz])
   for (int i = 0; i < sz; i++) {
      if (omp_is_initial_device()) {
         printf("Running on host\n");    
      } else {
         int nthreads= omp_get_num_threads();
         int nteams= omp_get_num_teams(); 
         printf("Running on device with %d teams (fixed) in total and %d threads in each team\n",nteams,nthreads);
      }
      fprintf(stdout, "Thread %d %i\n", omp_get_thread_num(), i );
      y[i] = a * x[i] + y[i];
   }
}
int main(int argc, char** argv) {
   float a = 2.0;
   int sz = 16;
   float *x = calloc( sz, sizeof *x );
   float *y = calloc( sz, sizeof *y );
   //Set values
   int num_devices = omp_get_num_devices();
   printf("Number of available devices %d\n", num_devices);
   saxpy( a, x, y, sz );
   return 0;
}

我试着用

gcc -O0 -fopenmp -foffload=nvptx-none -o mintest mintest.c

或者具有上述标志。

xlpyo6sf

xlpyo6sf1#

我猜问题是GCC无法处理GPU上运行的代码区域内的printf。GPU通常不擅长任何形式的I/O发生,因此您应该避免在卸载的代码区域内调用printfreadwrite等。
如果你想检测代码是在GPU设备上还是在主机上运行,那么你可以使用这样的模式:

void test_on_gpu(void) {
    int on_device = 0;
    #pragma omp target teams map(from:on_device)
    {
        #pragma omp parallel
        {
            #pragma omp master
            {
                if (0 == omp_get_team_num()) {
                    on_device = !omp_is_initial_device()
                }
            }
        }
    }
    printf("on GPU: %s\n", on_device ? "yes" : "no");
}

代码的作用是:

  • 转换到GPU设备(target
  • 在第一个OpenMP团队中使用一个线程(主线程master),并在那里使用并行区域
  • 确定执行是否发生在GPU上
  • 通过map(from:on_device)返回测试结果

相关问题