时间戳

首页 > 解决方案 > 从 CUDA 文件调用 Fortran OpenACC。如何用 PGI 编译？

cuda - 从 CUDA 文件调用 Fortran OpenACC。如何用 PGI 编译？

问题描述

我有一个 CUDA 代码，我想在其中包含由 Fortran 和 OpenACC 内核组成的外部代码。我有两个文件，其中包含以下内容，灵感来自 NVIDIA 网站上的讨论。文件main.cu如下：

#include <cstdio>

extern "C" void saxpy(int*, float*, float*, float*);

int main(int argc, char **argv)
{
    float* x;
    float* y;
    float* dx;
    float* dy;

    int n = 1<<20;

    x = (float*) malloc(n*sizeof(float));
    y = (float*) malloc(n*sizeof(float));

    for (int i=0; i<n; ++i)
    {
        x[i] = 1.f;
        y[i] = 0.f;
    }

    cudaMalloc((void**) &dx, (size_t) n*sizeof(float));
    cudaMalloc((void**) &dy, (size_t) n*sizeof(float));

    cudaMemcpy(dx, x, (size_t) n*sizeof(float), cudaMemcpyHostToDevice);
    cudaMemcpy(dy, y, (size_t) n*sizeof(float), cudaMemcpyHostToDevice);

    float factor = 2.f;
    saxpy(&n, &factor, dx, dy);

    cudaMemcpy(y, dy, (size_t) n*sizeof(float), cudaMemcpyDeviceToHost);
    printf("%f, %f\n", y[0], y[n-1]);

    return 0;
}

第二个文件saxpy.f90是：

subroutine saxpy(n, a, x, y) bind(c, name="saxpy")
    use iso_c_binding, only: c_int, c_float

    integer(kind=c_int), intent(in) :: n
    real(kind=c_float), intent(in) :: a
    real(kind=c_float), dimension(n), intent(in) :: x(n)
    real(kind=c_float), dimension(n), intent(inout) :: y(n)

    !$acc parallel deviceptr(x, y)
    do i = 1, n
        y(i) = y(i) + a*x(i)
    end do
    !$acc end parallel
end subroutine

我如何编译它nvcc并结合 PGI 编译器？我尝试了许多不同的选择，但我总是以未解决的外部问题结束。

我尝试的是：pgf90 -ta=tesla:cc35 -acc saxpy.f90 -c对于 Fortran 文件，它编译得很好。下一步是我卡住的地方。这：nvcc -arch=sm_35 -ccbin pgc++ main.cu saxpy.o产生未解决的外部问题，我不确定如何解决它。如何找出要包含的外部库？

标签： cudanvccopenaccpgi

解决方案

由于您没有将 OpenACC 或 Fortran 运行时库添加到链接中，因此很可能缺少这些符号。此外，当不使用 PGI 驱动程序链接时，您需要添加“nordc”标志。例如：

% pgfortran -c -ta=tesla:cc70,nordc saxpy.f90                                       
% nvcc -arch=sm_70 -ccbin pgc++ -Xcompiler "-ta=tesla:cc70 -pgf90libs" main.cu saxpy.o
% a.out
2.000000, 2.000000

不过，我建议使用 pgfortran 进行链接，这样您就可以使用 RDC 而无需添加 Fortran 运行时库：

% nvcc -arch=sm_70 -ccbin pgc++ -c main.cu
% pgfortran -Mcuda -ta=tesla:cc70 -Mnomain saxpy.f90 main.o
saxpy.f90:
% a.out
2.000000, 2.000000

推荐阅读