OPEN ACC - как управлять данными Struct внутри подпрограмм?

У меня есть такая структура:

typedef struct Data_{
  double **v;   
  .
  .
  .
  double *press; 
}Data;

В основной функции у меня есть цикл while, внутри которого он называется подпрограммой, в которой объявлены данные, и я использую следующую прагму:

static Data data;
#pragma acc enter data copyin(data[:7])

В этой программе он называется другим (RHS (&data, ...)). Опять же, внутри последнего я вызываю другую процедуру (RHS1(data,...)), в которой есть цикл, который я хочу ускорить:

#pragma acc parallel loop present(data[:7])
  for (i = beg; i <= end; i++) {
    rhs[i][MX1] += dt*data->src[i][MX1];
    . += .
    . += .
    . += .
    . += .
    . += .
    rhs[i][ENG] += dt*sweep->src[i][ENG];
  }

Я получаю эту ошибку при компиляции с помощью -managed: FATAL ERROR: переменная в разделе данных частично присутствует на устройстве: name = data


loops struct data-management openacc
person Smoden    schedule 17.07.2020    source источник

Ответы (1)


arrow_upward
0
arrow_downward

Пара вопросов.

#pragma acc enter data copyin(data[:7])

Это означает, что вы хотите скопировать массив из 7 структур данных, а не структуру из 7 элементов. Измените это на:

#pragma acc enter data copyin(data)

Компилятор знает размер структуры и выделяет соответствующий размер в копии устройства.

Однако предложения копирования выполняют только поверхностную копию структуры. Итак, здесь вы копируете указатели узлов элементов данных на устройство, и доступ к src приведет к ошибке недопустимого адреса. Таким образом, вам нужно будет сделать дополнительный шаг - вручную создать глубокую копию всех участников, которые будут использоваться на устройстве.

Итак, позже в программе, после того, как вы выделите src, добавьте дополнительную директиву:

#pragma acc enter data copyin(data.src[0:nx][0:ny])

(замените nx и ny на фактические переменные размеров)

Это сначала создаст массив src на устройстве, а затем подключит устройство src к копии данных устройства. attach в основном заполняет значение устройства data.src адресом массива src устройства.

Затем в параллельном цикле используйте только present (data).

Обратите внимание: если вы используете директиву update для синхронизации данных, обязательно обновляйте только src, а не data. Опять же, это будет выполнять неглубокую копию, поэтому при обновлении данных будут копироваться указатели хоста / устройства, что вызовет ошибки времени выполнения.

Я написал базовый пример того, как это сделать, в главе 5 книги Параллельное программирование с OpenACC с примером, который вы можете найти по адресу: https://github.com/rmfarber/ParallelProgrammingWithOpenACC/blob/master/Chapter05/array_of_structs.c

person Mat Colgrove    schedule 17.07.2020