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

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

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

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

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

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


openacc
person dwn    schedule 25.08.2016    source источник

Ответы (1)


arrow_upward
1
arrow_downward

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

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

    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.

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

person Mat Colgrove    schedule 25.08.2016