c++ - OpenACC 管理二维数据移动

Question

如何在主机 CPU 和 GPU 之间交换二维元素？我尝试将此 2d 元素用作 w[0:(n_hidden*i)-1]，但编译器反馈告诉我有问题

这是RBM算法函数:

double RBM::propdown(int *h, int i, double b) {
#pragma acc data region \
copyin(w[0:(n_hidden*i)-1],h[0:n_hidden],b) create(pre_sigmoid_activation)
double pre_sigmoid_activation = 0.0;
#pragma acc parallel loop reduction(+:pre_sigmoid_activation) 
for(int j=0; j<n_hidden; j++) {
pre_sigmoid_activation += W[j][i] * h[j];
}
pre_sigmoid_activation += b;  
return sigmoid(pre_sigmoid_activation);
#pragma acc exit data \
delete ( pre_sigmoid_activation)
}

Answer 1

不需要线性化数组。只需使用多个括号。

#pragma acc data copyin(W[0:n_hidden][0:N])

我还看到了许多其他问题。

数据指令没有“区域”子句。您可能会将其与“输入数据”子句或作为 OpenACC 基础的 PGI 加速器模型混淆。

不需要将“b”放在数据子句中，因为它实际上并未在计算区域中使用。此外，通过将它放在数据子句中，您可以使“b”成为设备上的全局引用。最好在数据子句中保留只读标量，以便将值作为参数传递，而不是需要从全局内存中获取它。

再次通过将标量变量“pre_sigmoid_activation”放入数据子句中，您创建了一个全局变量。这里，减少的结果将存储在这个设备变量中，并且不会在主机上自动更新。为此，您需要添加一个“更新”指令。更好的是，只需将其从数据子句中删除，归约结果将更新到主变量。

您有一个不匹配的“退出数据”指令(应该有一个相应的“输入数据”指令)。此外，该指令位于 return 语句之后，因此永远不会执行，从而将数据留在设备上。

最后，由于 C++ 区分大小写，请确保 OpenACC 指令中的变量名称与实际变量名称匹配。即“W”而不是“w”。

下面是我将如何编写循环。请注意，我不知道“W”的第二个维度的大小，所以只使用了“N”。请相应更新。

#pragma acc data copyin(W[0:n_hidden][0:N],h[0:n_hidden])
{
  #pragma acc parallel loop reduction(+:pre_sigmoid_activation)
  for(int j=0; j<n_hidden; j++) {
    pre_sigmoid_activation += W[j][i] * h[j];
  }
}