OpenACC и объектно-ориентированный C++

Question

Я пытаюсь написать объектно-ориентированный код на C++, который распараллелен OpenACC. Мне удалось найти некоторые вопросы о стеке и вопросы GTC на OpenACC, но я не смог найти примеры реальных объектов объектно-ориентированного кода.

В 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 является плоской структурой, которая сама по себе обрабатывает указатели, она должна работать, чтобы скопировать ее в качестве члена? Или нужно быть указателем на структуру, и указатели должны быть связаны с директивами? Тогда я боюсь проблемы, что я должен использовать указатели псевдонимов, как предложил Джефф Ларкин в the above mentioned question. Я не прочь сделать работу, чтобы запустить ее, но я не могу найти ссылку, как это сделать. Использование директив компилятора keepgpu,keepptx помогает немного понять, что делает компилятор, но я предпочел бы альтернативу обратному инженерному генерированному ptx-коду.

Любые указатели на полезный справочный проект или документы высоко оценены.

Answer 1

В заголовке OpenACCArray1.h удалите два "#pragma acc введите данные create (this)" прагмы. Что происходит, так это то, что конструктор «Data» создает объекты «a» и «a3» на устройстве. Следовательно, когда вторая область ввода вводится в конструкторе OpenACCArray, это указатель уже существует.

Он работает, когда имеется только один элемент данных, так как «a3» и «Data» используют один и тот же адрес для этого указателя. Следовательно, когда появляется вторая входная прагма данных, настоящая проверка видит, что она уже на устройстве, поэтому не создала ее снова. Когда добавляется «a», размер «Данные» в два раза превышает размер «a», поэтому в текущей проверке видно, что этот указатель уже существует, но имеет другой размер, чем раньше. Вот что означает «частично присутствующая» ошибка. Данные есть, но имеют иной, чем ожидалось, размер.

Только родительский класс/структура должен создать этот указатель на устройстве.

Надеется, что это помогает, Mat