Почему следующий код:
#include <iostream>
int main(int argc, char const *argv[])
{
int sum = 0;
int *array;
array = new int [100];
#pragma acc enter data create(array[0:100],sum)
#pragma acc parallel loop present(array[0:100])
for (int i = 0; i < 100; ++i)
{
array[i] = 1;
}
#pragma acc parallel loop present(array[0:100],sum) reduction(+:sum)
for (int i = 0; i < 100; ++i)
{
sum += array[i];
}
#pragma acc exit data delete(array[0:100]) copyout(sum)
std::cout << sum << std::endl;
return 0;
}
Возвращать разные результаты при каждом выполнении?
$ pgcpp -acc -Minfo main.cpp
main:
7, Generating enter data create(sum)
Generating enter data create(array[:100])
Generating present(array[:100])
Accelerator kernel generated
12, #pragma acc loop gang, vector(256) /* blockIdx.x threadIdx.x */
7, Generating Tesla code
15, Generating present(array[:100])
Generating present(sum)
Accelerator kernel generated
18, #pragma acc loop gang, vector(256) /* blockIdx.x threadIdx.x */
20, Sum reduction generated for sum
15, Generating Tesla code
25, Generating exit data copyout(sum)
Generating exit data delete(array[:100])
$ ./a.out
100
$ ./a.out
200
$ ./a.out
300
$ ./a.out
400
Согласно стандарту OpenACC:
В директиве exit data данные копируются обратно в локальную память и
перераспределена.
Казалось бы, что sum
является не освобождается и используется повторно (и увеличивается) при каждом запуске программы. Кроме того, +
оператор в reduction
Директива инициализирует редукционную переменную в 0
так что это не должно происходить, даже если sum
не были распределены между казнями.
Я могу избежать этого поведения, используя copyin
вместо create
за sum
в enter data
директива или настройка sum = 0
в одной банде, в одном рабочем ядре:
#pragma acc parallel present(sum) num_gangs(1) num_workers(1)
sum = 0;
Но это неудовлетворительно, так как требует либо дорогостоящего хоста для копирования данных устройства, либо соответственно запуска ядра. Почему моя программа ведет себя так?
Вы неверно истолковываете значение инициализации оператора редукции. Ссылаясь на спецификация openACC, С. 20-21:
Предложение сокращения допускается для параллельной конструкции. Он определяет оператор редукции и одну или несколько скалярных переменных. Для каждой переменной создается личная копия для каждой параллельной банды и инициализируется для этого оператора. В конце региона значения для
каждая банда объединяется с использованием оператора сокращения, и результат в сочетании со значением исходной переменной и сохраняется в исходной переменной.
Это означает, что общая проблема сокращения разбита на части, каждая часть обрабатывается бандой. Часть проблемы, решаемой бандой, будет использовать указанное значение инициализации для редукционной переменной. Однако, когда будет создан конечный результат, отдельные результаты каждой банды будут объединены со значением исходной переменной (sum
в вашем случае), и это будет результатом.
Таким образом, вы должны правильно инициализировать sum
, возможно, используя один из методов, которые вы изложите в своем вопросе.
Кроме того, хотя это не суть проблемы, обратите внимание, что ни освобождение, ни распределение не влияют на содержимое памяти. Новая переменная, расположенная в этом месте, без надлежащей инициализации, в данный момент получит значение в этом месте.