Как скопировать на графический процессор вектор-указатель-вектор-память, выделенный в OpenACC - PullRequest
Новая
       62

Как скопировать на графический процессор вектор-указатель-вектор-память, выделенный в OpenACC

0 голосов
/ 20 декабря 2018

Я пытаюсь скопировать на графический процессор вектор векторный массив.

Я пытался с предложением OpenACC copyin.Предложение copyin не копирует все базовые данные моего массива.Когда я пытаюсь получить доступ к базовым векторным данным, я получаю ошибку «недопустимый доступ по адресу» во время выполнения. 

vector<int32_t> *k1p = new vector<int32_t>[bin_num];
for (int i = 0; i < bin_mum; i++) {
  //......
  k1p[i].push_back(i);
}
#pragma acc kernels loop independent copyin(k1p[0:bin_num])
for (int i = 0; i < bin_mum; i++) {
  //........
  for (vector<int32_t>::const_iterator i2_it=k1p[i].begin(); i2_it!=k1p[i].end(); i2_it++) {
    //.......
  }
  //..........
}

Я хочу получить доступ к элементу базового вектора k1p [i], но на самом деле этот код компилируетсяс компилятором pgi, но когда я запускаю этот код, я получаю

вызов cuStreamSynchronize, возвращаемую ошибку 700: недопустимый адрес во время выполнения ядра

Ответы [ 2 ]

0 голосов
/ 03 января 2019

Опять же, я настоятельно рекомендую использовать унифицированную память CUDA (управляемую) при использовании векторов C ++ на устройстве.Векторы представляют собой контейнерный тип с тремя частными указателями.Если копия или обновление OpenACC делает поверхностную копию, при помещении вектора в область данных вы будете копировать указатели, а не данные, на которые они указывают.Хуже того, поскольку указатели являются частными, они не могут быть обновлены с помощью указателя устройства.

Вместо этого вам потребуется создать массив временных указателей для теневого копирования данных векторного массива, а затем использовать эту временную переменную.на устройстве.Что-то вроде следующего: 

% cat vector.data.cpp
#include <iostream>
#include <vector>
#include <cstdint>

using namespace std;

int main() {
   const int bin_num = 1024;
   long sum = 0;

   vector<int32_t> *k1p = new vector<int32_t>[bin_num];
   for (int i = 0; i < bin_num; i++) {
       k1p[i].push_back(i);
   }

   int32_t ** temp = new int32_t*[bin_num];
   int * sizes = new int[bin_num];
   #pragma acc enter data create(temp[0:bin_num][0:0])
   for (int i = 0; i < bin_num; i++) {
      int sze = k1p[i].size();
      sizes[i] = sze;
      temp[i] = k1p[i].data();
      #pragma acc enter data copyin(temp[i:1][:sze])
   }
   #pragma acc enter data copyin(sizes[:bin_num])

  #pragma acc parallel loop gang vector reduction(+:sum) present(temp,sizes)
  for (int i = 0; i < bin_num; i++) {
     for (int j=0; j< sizes[i]; ++j) {
         sum += temp[i][j];
     }
  }
  cout << "Sum: " << sum << endl;
  #pragma acc exit data delete(sizes)
  #pragma acc exit data delete(temp)
  delete [] sizes;
  delete [] k1p;
}

% pgc++ vector.data.cpp --c++11 -ta=tesla  -V18.10
% a.out
Sum: 523776

При использовании управляемой памяти адреса находятся в едином пространстве памяти, доступном как для хоста, так и для устройства.Следовательно, когда вы открываете «начало» и «конец», адреса, которые они возвращают на устройстве, действительны.Например: 

% cat vector.cpp
#include <iostream>
#include <vector>
#include <cstdint>
using namespace std;

int main() {
   const int bin_num = 1024;
   long sum = 0;
   vector<int32_t>::const_iterator i2_it;

   vector<int32_t> *k1p = new vector<int32_t>[bin_num];
   for (int i = 0; i < bin_num; i++) {
       k1p[i].push_back(i);
   }

#pragma acc parallel loop reduction(+:sum) private(i2_it)
  for (int i = 0; i < bin_num; i++) {
    for (i2_it=k1p[i].begin(); i2_it!=k1p[i].end(); i2_it++) {
         sum += *i2_it;
     }
  }
  cout << "Sum: " << sum << endl;
}

% pgc++ vector.cpp --c++11 -ta=tesla:managed -V18.10
% a.out
Sum: 523776

Другой вариант - написать свой собственный вектор, подобный классу.Я написал основной пример того, как это сделать, как часть главы 5 Параллельное программирование с OpenACC .Образцы кода можно найти по адресу https://github.com/rmfarber/ParallelProgrammingWithOpenACC

0 голосов
/ 20 декабря 2018

Предложения данных OpenACC выполняют только поверхностные копии объекта.Поскольку «вектор» представляет собой набор из трех указателей, это означает, что помещение вектора в предложение copyin будет копировать только указатели, а не данные, на которые он указывает.

Если вы используете PGI, самый простой способнужно использовать унифицированную память CUDA (т. е. добавить флаг «-ta = tesla: managed») и заставить среду выполнения CUDA управлять перемещением данных за вас.

В противном случае вам необходимо выполнить углубленное руководствокопия вектора.Это может быть немного сложно, особенно с векторами, поэтому дайте мне знать, если вам нужен пример.

...