OpenACC数据移动
OpenACC数据移动
提问于 2016-06-02 16:07:51
我对OpenACC非常陌生，我不太了解数据移动和“#语用acc数据”子句。
我有一个用C编写的程序，代码的摘录是这样的：
#pragma acc data create(intersectionSet[0:intersectionsCount][0:4]) // line 122
#pragma acc kernels // line 123
for (int i = 0; i<intersectionsCount; i++){ // line 124
intersectionSet[i][0] = 9; // line 125
}intersectionsCount的值为210395。编译并运行上述代码后，请执行以下操作：
pgcc -o rect_openacc -fast -Minfo -acc -ta=nvidia,time rect.c我有这样的输出：
time(us): 1,475,607
122: data region reached 1 time
31: kernel launched 210395 times
grid: [1] block: [128]
device time(us): total=1,475,315 max=15 min=7 avg=7
elapsed time(us): total=5,451,647 max=24,028 min=24 avg=25
123: compute region reached 1 time
124: kernel launched 1 time
grid: [1644] block: [128]
device time(us): total=292 max=292 min=292 avg=292
elapsed time(us): total=312 max=312 min=312 avg=312
156: data region reached 1 time在阅读了输出后，我有一些问题：
- 我不知道为什么它说第31行，因为第31行没有acc语用。是不是意味着我无法追踪到什么？
- 在"31:内核启动210395次“中，它说它启动了内核的210395倍。我不知道内核需要启动这么多次是否正常，因为这个部分花费了5,451,647(我们)，我认为它有点长。我认为for-循环很简单，不应该花那么多时间。我是不是用错了语用方法？
更新
我确实有几个程序的头文件。但是这些文件没有"acc数据“或"acc内核”的标记。
使用"-Minfo=all“编译代码后，结果如下：
breakStringToCharArray:
11, include "stringHelper.h"
50, Loop not vectorized/parallelized: contains call
countChar:
11, include "stringHelper.h"
74, Loop not vectorized/parallelized: not countable
extractCharToIntRequiredInt:
11, include "stringHelper.h"
93, Loop not vectorized/parallelized: contains call
extractArray:
12, include "fileHelper.h"
49, Loop not vectorized/parallelized: contains call
isRectOverlap:
13, include "shapeHelper.h"
23, Generating acc routine vector
Generating Tesla code
getRectIntersection:
13, include "shapeHelper.h"
45, Generating acc routine vector
Generating Tesla code
getRectIntersectionInGPU:
13, include "shapeHelper.h"
69, Generating acc routine vector
Generating Tesla code
max:
13, include "shapeHelper.h"
98, Generating acc routine vector
Generating Tesla code
min:
13, include "shapeHelper.h"
118, Generating acc routine vector
Generating Tesla code
main:
64, Loop not vectorized/parallelized: contains call
108, Loop not vectorized/parallelized: contains call
122, Generating create(intersectionSet[:intersectionsCount][:4])
124, Loop is parallelizable
Accelerator kernel generated
Generating Tesla code
124, #pragma acc loop gang, vector(128) /* blockIdx.x threadIdx.x */我以这种方式创建intersectionSet：
intersectionSet = (int **)malloc(sizeof(int **) * intersectionsCount);
for (i = 0; i<intersectionsCount; i++){
intersectionSet[i] = (int *)malloc(sizeof(int *) * 4);
}回答 1
Stack Overflow用户
回答已采纳
发布于 2016-06-03 16:31:22
因为您有指向指针数组的指针，"**"，(至少我猜这就是intersectionSet )，编译器必须首先将指针分配给设备上的指针，然后循环遍历每个元素来分配各个设备数组。最后，它需要启动一个内核来设置设备上的指针值。这里有一些伪代码来帮助说明。
devPtrPtr = deviceMalloc(numElements*pointer size);
for (i=0; i < numElements; ++i) {
devPtr = deviceMalloc(elementSize * dataTypeSize);
call deviceKernelToSetPointer<<<1,128>>(devPtrPtr[i],devPtr);
}为了帮助您编写代码，我会切换列长度4和行长度"intersectionsCount“的维度。这也将有助于设备上的数据访问，因为“向量”循环应该与步长-1(连续)维相对应，以避免内存发散。
希望这能帮上忙
垫子
页面原文内容由Stack Overflow提供。腾讯云小微IT领域专用引擎提供翻译支持
原文链接：
https://stackoverflow.com/questions/37596968
复制相关文章
相似问题
腾讯云开发者
