打开 ACC：PGI 编译器没有输出答案

【问题标题】：open ACC : no output with PGI compiler打开 ACC：PGI 编译器没有输出
【发布时间】：2014-02-28 02:19:40
【问题描述】：

我是一名本科生，正在实习中学习开放式 ACC。

我自己制作了基本的示例代码，用于测试 PGI 工作站编译器（14.1 版）运行良好，它与命令行指令 'pgcc -acc lab04.c -Minfo' 配合使用

但是我有一个奇怪的问题。 @@ 任何地方都没有输出！！ @@

我确信编译器运行良好，内核也是如此。但我猜肯定有错误。

你们能帮帮我吗？

我将发布整个代码。

#include <stdio.h>
#include <stdlib.h>


float scaled(float* v1, float* v2, float a, int n)
{
int i;
float sum = 0.0f;

#pragma acc kernels loop
for(i=0;i<n;i++)
{
    v1[i]+=a*v2[i];
    sum+=v1[i];
}

return sum;
}

int main(int argc, char* argv[])
{
int n;

float *vector1;
float *vector2;

if( argc > 1 )
    n = atoi( argv[1] );
else
    n = 100000;
if( n <= 0 ) n = 100000;


vector1=(float*)malloc(n*sizeof(float));
vector2=(float*)malloc(n*sizeof(float));


scaled(vector1, vector2, 3.3, n);


printf("programming done\n");

return 0;

}

【问题讨论】：

“无输出”是什么意思？您应该看到的唯一输出是“programming done”，这是代码中唯一的 printf 语句。

标签： gpu hpc openacc

【解决方案1】：

仅供参考，我能够在 PGI 用户论坛上帮助该用户（请参阅：http://www.pgroup.com/userforum/viewtopic.php?t=4164）。除了 Robert 指出的指针混叠问题外，他使用的是较旧的 NVIDIA GPU，默认情况下该 GPU 不再是目标，需要添加“-ta=tesla:cc10”。

请注意，“特斯拉”目标是在 14.1 中添加的。对于早期版本，请将“tesla”替换为“nvidia”。

【讨论】：

【解决方案2】：

我尝试使用 PGI 13.10 编译器工具通过以下命令行编译您的程序：

pgcc -acc -ta=nvidia,cc20,cuda5.0 -o t1 t1.c -Minfo

得到这个输出：

scaled:
 10, Generating present_or_copy(v1[0:n])
     Generating present_or_copyin(v2[0:n])
     Generating NVIDIA code
     Generating compute capability 2.0 binary
 11, Complex loop carried dependence of '*(v2)' prevents parallelization
     Loop carried dependence of '*(v1)' prevents parallelization
     Loop carried backward dependence of '*(v1)' prevents vectorization
     Accelerator scalar kernel generated
 14, Sum reduction generated for sum

虽然这确实表明编译“成功”，但有关“防止并行化”的消息表明编译器在真正利用加速器方面成功。当您看到消息Accelerator scalar kernel generated 时，您通常会对结果不满意。

当我运行上面编译的程序时，我确实收到了运行时错误：

call to cuLaunchKernel returned error 701: Launch out of resources

这是来自 CUDA 运行时子系统的错误。根据您尝试运行的加速器设备类型，您可能会看到也可能不会看到此类错误。我们可以深入研究如何解决这个问题，但这真的是无关紧要的，因为您的程序结构不正确，无法充分利用加速器。

编译器发出“防止并行化”消息，因为它对指针vector1（或v1）和vector2（或v2）非常严格。它假定这些指针可以相互重叠，因此在这种情况下无法创建正确的并行程序。由于这可能不是您的意图（您可能希望 v1 和 v2 引用单独的空格），您可以通过使用 C99 restrict 关键字修改 scaled 函数参数来“放心”编译器。这允许编译器按照您可能想要的方式完成其工作。

这是修改后的代码和结果：

$ cat t1.c
#include <stdio.h>
#include <stdlib.h>

float scaled(float *restrict v1, float *restrict v2, float a, int n)
{
  int i;
  float sum = 0.0f;

  #pragma acc kernels
  for(i=0;i<n;i++)
  {
    v1[i]+=a*v2[i];
    sum+=v1[i];
  }

  return sum;
}

int main(int argc, char* argv[])
{
  int n;

  float *vector1;
  float *vector2;

  if( argc > 1 )
    n = atoi( argv[1] );
  else
    n = 100000;
  if( n <= 0 ) n = 100000;


  vector1=(float*)malloc(n*sizeof(float));
  vector2=(float*)malloc(n*sizeof(float));


  scaled(vector1, vector2, 3.3, n);


  printf("programming done\n");

  return 0;

}

$ pgcc -acc -ta=nvidia,cc20,cuda5.0 -o t1 t1.c -Minfo
scaled:
     10, Generating present_or_copy(v1[0:n])
         Generating present_or_copyin(v2[0:n])
         Generating NVIDIA code
         Generating compute capability 2.0 binary
     11, Loop is parallelizable
         Accelerator kernel generated
         11, #pragma acc loop gang, vector(128) /* blockIdx.x threadIdx.x */
     14, Sum reduction generated for sum
$ ./t1
programming done
$

如果您无法获得这样的结果，则可能是您尝试运行的机器/工具存在问题。

【讨论】：