多维数组拷贝OpenACC答案

【问题标题】：Multi-dimensional array copy OpenACC多维数组拷贝OpenACC
【发布时间】：2012-10-04 17:06:41
【问题描述】：

我有一个 2D 矩阵 SIZE x SIZE，我正在尝试将其复制到 GPU。

我是这样分配矩阵的：

#define SIZE 1024
float (*a)(SIZE) = (float(*)[SIZE]) malloc(SIZE * SIZE * sizeof(float));

我的 ACC 地区有这个：

void mmul_acc(restrict float a[][SIZE],
              restrict float b[][SIZE],
              restrict float c[][SIZE]) {
#pragma acc data copyin(a[0:SIZE][0:SIZE], b[0:SIZE][0:SIZE]) \
    copyout c[0:SIZE][0:SIZE])
{
  ... code here...
}

使用 PGI 编译器进行编译时，使用 -Minfo=acc，编译器会告诉我：

Generating copyin(a[0:1024][0:])

[0:1024][0:] 是什么意思？为什么不是 [0:1024][0:1024] ???

如果我不声明矩阵，而是声明大小为 SIZE*SIZE 的数组，则这样做

#pragma acc copyin(a[0:SIZE*SIZE])

生成以下编译器消息

Generating copyin(a[0:16777216])

代码实际上以相同的方式工作，相同的性能，相同的结果。

显然，编译器在这两种方式中都生成了相同的代码，但应该是这样，但消息并不直接。

我在 Linux64 机器上使用 PGI 加速器 12.8。我正在使用 -Minfo=acc

进行编译

注意：这个问题已经过编辑，现在它并没有多大意义，但它可能对更多人有用。

【问题讨论】：

我假设您使用的是 PGI OpenACC 编译器。您是否使用 -Minfo 开关进行编译？在 C 中，当将双下标数组作为参数传递给函数时，需要确定第一个下标的范围。因此，“Generating copyin”消息只是编译器确认它正在双下标数组 a 上创建主机->设备副本，第一个下标范围为 1024。大概 SIZE 设置为 1024。回答第二个问题关于为什么代码运行速度慢 10 倍，在这两种情况下使用 -Minfo 开关查看完整的编译器输出会很有用。
另外，float (*a)(SIZE) 在语法上是否有效？你为什么不直接做float **a = (float**)malloc(SIZE * SIZE * sizeof(float));？
罗伯特：是的，我是。我刚刚编辑了问题以添加更多信息。大约慢了 10 倍，我犯了一个错误，我已经纠正了这个问题。谢谢
harrism：是的，它告诉编译器分配一个一维数组，但将其解释为一个矩阵（基本上你有两个索引访问它的好处，但它保证是连续分配的. 谢谢！
我应该在这篇文章中补充一点，编译器消息 Generating copyin(a[0:1024][0:]) 已在 PGI 加速器 12.9.0 中修复，并提供 Generating copyin(a[0:1024][0:1024])

标签： cuda gpgpu openacc

【解决方案1】：

此问题已在最新的 PGI 编译器 12.9.0 中得到修复。编译器现在返回以下消息：

Generating copyin(a[0:1024][0:1024])

【讨论】：