OpenMP GPU 卸载数学库？答案

【问题标题】：OpenMP GPU offloading math library?OpenMP GPU 卸载数学库？
【发布时间】：2018-09-06 22:59:53
【问题描述】：

我正在尝试使用 OpenMP 4+ 指令卸载 GPU 代码。我正在使用带有 GCC 7.2 的 ubuntu 16.04，对于一般情况，它工作正常。当我试图卸载调用“math.h”中定义的 sqrtf 函数的代码时，我的问题就出现了。麻烦的代码是这样的：

#pragma omp target teams distribute \
map(to:posx[:n],posy[:n],posz[:n]) \
map(from:frcx[:n],frcy[:n],frcz[:n])
for (int i = 0; i < n; i++) {
  frcx[i] = 0.0f;
  frcy[i] = 0.0f;
  frcz[i] = 0.0f;

  for (int j = 0; j < n; j++) {
    float dx = posx[j] - posx[i];
    float dy = posy[j] - posy[i];
    float dz = posz[j] - posz[i];
    float distSqr = dx*dx + dy*dy + dz*dz + SOFTENING;
    float invDist = 1.0f / sqrtf(distSqr);
    float invDist3 = invDist * invDist * invDist;

    frcx[i] += dx * invDist3;
    frcy[i] += dy * invDist3;
    frcz[i] += dz * invDist3;
  }
}

当我尝试编译它时：

$ gcc -Wall -O2 -march=native -mtune=native -fopenmp -o nbody_cpu_arrays_parallel_gpu common_funcs.c nbody_cpu_arrays_parallel_gpu.c -lm
unresolved symbol sqrtf
collect2: error: ld returned 1 exit status
mkoffload: fatal error: x86_64-linux-gnu-accel-nvptx-none-gcc-7 returned 1 exit status
compilation terminated.
lto-wrapper: fatal error: /usr/lib/gcc/x86_64-linux-gnu/7//accel/nvptx-none/mkoffload returned 1 exit status
compilation terminated.
/usr/bin/ld: error: lto-wrapper failed
collect2: error: ld returned 1 exit status

在将 OMP 代码卸载到 GPU 时，如何使用平方根运算（或其他数学函数）？

【问题讨论】：

您是否尝试过查找是否有可用的 GPU 数学库？例如，Nvidia provides these functions in their math library。我也会尝试链接到该库，因为 libm 仅适用于您的主机 CPU 架构。
对于这种特殊情况，-ffast-math 有帮助吗？

标签： c math gcc gpu openmp

【解决方案1】：

我遇到了类似的问题。 https://github.com/bisqwit/cpp_parallelization_examples/blob/master/README.md 非常有帮助地描述了解决方案：

卸载时，如果出现以下情况，您可能会从数学函数中遇到链接器问题你做了一个优化的构建。要解决，请添加 -foffload=-lm -fno-fast-math -fno-associative-math

作为参考，我用 sqrt 得到的错误：

libgomp: Link error log ptxas application ptx input, line 138; error   : Label expected for argument 0 of instruction 'call'
ptxas application ptx input, line 138; fatal   : Call target not recognized
ptxas <macro util>, line 9; error   : Illegal modifier '.div' for instruction 'mov'
ptxas fatal   : Ptx assembly aborted due to errors


libgomp: cuLinkAddData (ptx_code) error: a PTX JIT compilation failed

libgomp: Cannot map target functions or variables (expected 2, have 4294967295)

还有 sqrtf：

unresolved symbol sqrtf
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: gcc/x86_64-pc-linux-gnu/7.3.0//accel/nvptx-none/mkoffload returned 1 exit status
compilation terminated.
/usr/bin/ld: error: lto-wrapper failed

【讨论】：

【解决方案2】：

clang 9.0 现在具有将标准数学库函数替换为等效版本的 ptx 代码 (nvidia gpu) 的功能，gcc 9.0 尚不支持该功能。

编译运行：https://www.hahnjo.de/blog/2018/10/08/clang-7.0-openmp-offloading-nvidia.html

clang 的提交：https://reviews.llvm.org/D61399

【讨论】：