【发布时间】:2016-12-29 23:18:59
【问题描述】:
我正在尝试编写与 OpenACC 并行的面向对象的 C++ 代码。 我能够找到一些关于 OpenACC 的 stackoverflow 问题和 GTC 演讲,但我找不到一些面向对象代码的真实示例。
在this question 中,显示了OpenACCArray 的示例,它在后台进行一些内存管理(代码在http://www.pgroup.com/lit/samples/gtc15_S5233.tar 上可用)。
但是,我想知道是否可以创建一个在更高级别上管理数组的类。例如
struct Data
{
// OpenACCArray<float> a;
OpenACCArray<Vector3<float>> a3;
Data(size_t len) {
#pragma acc enter data copyin(this)
// a.resize(len);
a3.resize(len);
}
~Data() {
#pragma acc exit data delete(this)
}
void update_device() {
// a.update_device();
a3.update_device();
}
void update_host() {
// a.update_host();
a3.update_host();
}
};
int main(int argc, char *argv[])
{
const size_t len = 32*128;
Data d(len);
d.update_device();
#pragma acc kernels loop independent present(d)
for (int i=0; i < len; ++i) {
float val = (float)i/(float)len;
d.a3[i].x = val;
d.a3[i].y = i;
d.a3[i].z = d.a3[i].x / d.a3[i].y;
}
d.update_host();
for (int i=0; i < len/128; ++i) {
cout << i << ": " << d.a3[i].x << "," << d.a3[i].y << "," << d.a3[i].z << endl;
}
cout << endl;
return 0;
}
有趣的是,这个程序可以工作,但是一旦我取消注释OpenACCArray<float> a;,即向该数据结构添加另一个成员,我就会出现内存错误。
FATAL ERROR: variable in data clause is partially present on the device.
由于OpenACCArray 结构是一个平面结构,它自己处理指针间接寻址,它应该可以将其复制为成员吗?
或者是否需要成为指向结构的指针并且指针必须与指令硬连线?
然后我担心我必须使用 jeff larkin 在the above mentioned question 建议的别名指针的问题。
我不介意做这项工作来让它运行,但我找不到任何参考如何做到这一点。
使用编译器指令 keepgpu,keepptx 有助于理解编译器在做什么,但我更喜欢反向工程生成的 ptx 代码的替代方法。
非常感谢任何指向有用的参考项目或文档的指针。
【问题讨论】:
-
您在该示例压缩包中使用 OpenACCArray 上的哪个版本?
-
另外,您能否包括您对 Vector3 的定义?我猜测了一下,但想确认我们正在构建相同的东西。
-
@jefflarkin 感谢您的帮助。我使用的是自定义类,但是将代码更改为使用 tarball 中示例 2 中的 float3。这是代码gist.github.com/danielwinkler/12ab5b73221faca89d69d83d72c633b7的要点
-
OpenACCArray也来自示例 2。 -
@jefflarkin 你能用提供的代码重现这个行为吗?