OpenACCArray функция подкачки

Question

Пытаясь создать объектно-ориентированную реализацию OpenACC, я наткнулся на этот вопрос.

Оттуда я взял код, предоставленный @mat-colgrove на GTC15 (код доступен по адресу http://www.pgroup.com/lit/samples/gtc15_S5233.tar).

Поскольку меня интересует, как использовать объекты для управления данными в OpenACC, я опубликовал еще один вопрос. Я был очень впечатлен легкостью OpenACCArray::swap функция, поэтому я создал небольшой пример, чтобы проверить его (см. суть).

• Сначала я попытался просто поменяться местами и надеюсь, что достаточно поменять местами указатели на хосте, но это заканчивается фатальной ошибкой памяти. (предположительно потому, что элементы размера и емкости не обновляются на устройстве)
• Я предположил, что более безопасный подход - это обновить хост, своп-массивы и обновить устройство. Это работает, но создает неправильные результаты.

Я собираю для nvidia ускорителей.

Answer 1

Похоже, это моя вина, так как я не проверял процедуру обмена.

Проблема здесь заключается в том, что пока код обменивает данные на хосте, копии объектов устройства все еще указывают на старый массив. Исправление состоит в том, чтобы повторно присоединить (т.е. установить указатели устройства объекта на правильные массивы) списки.

    void swap(OpenACCArray<type>& x)
    {
        type* tmp_list = list;
        int tmp_size = _size;
        int tmp_capacity = _capacity;
        list = x.list;
        _size = x._size;
        _capacity = x._capacity;
        x.list = tmp_list;
        x._size = tmp_size;
        x._capacity = tmp_capacity;
#ifdef _OPENACC
#pragma acc update device(_size,_capacity,x._size,x._capacity)
        acc_attach((void**)&list);
        acc_attach((void**)&x.list);
#endif

    }

"acc_attach" - это расширение PGI, которое, будем надеяться, будет принято в стандарте OpenACC 3.0.

Спасибо за попытку и дайте мне знать, если у вас возникнут другие проблемы. - коврик