OpenACC: сложный цикл, переносящий зависимость '*(*(b))', предотвращает распараллеливание
Я использую OpenACC с динамическим распределением массива. Вот как я выделяю:
float **a;
float **b;
float **c;
float **seq;
a=(float**)malloc(SIZE*sizeof(float*));
b=(float**)malloc(SIZE*sizeof(float*));
c=(float**)malloc(SIZE*sizeof(float*));
seq=(float**)malloc(SIZE*sizeof(float*));
for(i=0; i<SIZE; i++){
a[i]=(float*)malloc(SIZE*sizeof(float));
b[i]=(float*)malloc(SIZE*sizeof(float));
c[i]=(float*)malloc(SIZE*sizeof(float));
seq[i]=(float*)malloc(SIZE*sizeof(float));
}
и вот как я распараллеливаю матрицу добавить:
#pragma acc kernels copyin(a[0:SIZE][0:SIZE],b[0:SIZE][0:SIZE]) copy(c[0:SIZE][0:SIZE])
for (i = 0; i < SIZE; ++i) {
for (j = 0; j < SIZE; ++j) {
c[i][j] = a[i][j] + b[i][j];
}
}
Когда я компилирую этот код с pgcc
это обнаружить зависимость от float**
указывает на итерации цикла и генерирует все скалярное ядро (1 блок 1 поток на блок), которое работает плохо, как ожидалось:
40, Complex loop carried dependence of '*(*(b))' prevents parallelization
Complex loop carried dependence of '*(*(a))' prevents parallelization
Complex loop carried dependence of '*(*(c))' prevents parallelization
Accelerator scalar kernel generated
CC 1.0 : 11 registers; 40 shared, 4 constant, 0 local memory bytes
CC 2.0 : 22 registers; 0 shared, 56 constant, 0 local memory bytes
Цикл, очевидно, параллелен, и я думаю, что это также может быть обнаружено компилятором. Мне интересно, как это объяснить pgcc
?
Заранее спасибо.
1 ответ
Решение
Я думаю, что нашел ответ. Ключ должен использовать independent
пункт:
#pragma acc data copyin(a[0:SIZE][0:SIZE],b[0:SIZE][0:SIZE]) copy(c[0:SIZE][0:SIZE])
{
# pragma acc region
{
#pragma acc loop independent vector(16)
for (i = 0; i < SIZE; ++i) {
#pragma acc loop independent vector(16)
for (j = 0; j < SIZE; ++j) {
c[i][j] = a[i][j] + b[i][j];
}
}
}
}