Последовательная работа в реализации GPU

Question

Я должен реализовать следующий алгоритм в GPU

for(int I = 0; I < 1000; I++){
    VAR1[I+1] = VAR1[I] + VAR2[2*K+(I-1)];//K is a constant
}

Каждая итерация зависит от предыдущей, поэтому распараллеливание затруднено. Я не уверен, что атомарная операция действительна здесь. Что я могу сделать?

РЕДАКТИРОВАТЬ:

VAR1 и VAR2 оба являются одномерным массивом.

VAR1[0] = 1

Answer 1

Это относится к категории проблем, называемых рекуррентными отношениями. В зависимости от структуры рекуррентного отношения, могут существовать решения в замкнутой форме, которые описывают, как вычислять каждый элемент индивидуально (т.е. параллельно, без рекурсии). Одной из ранних основополагающих работ (по параллельным вычислениям) были Когге и Стоун, и существуют рецепты и стратегии для распараллеливания конкретных форм.

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

В вашем случае, давайте посмотрим, сможем ли мы найти что-нибудь, просто наметив, что первые несколько терминов VAR1 должен выглядеть так, подставляя предыдущие термины в новые:

i      VAR1[i]
___________________
0        1
1        1 + VAR2[2K-1]
2        1 + VAR2[2K-1] + VAR2[2K]
3        1 + VAR2[2K-1] + VAR2[2K] + VAR2[2K+1]
4        1 + VAR2[2K-1] + VAR2[2K] + VAR2[2K+1] + VAR2[2K+2]
...

Надеюсь, что на вас выскакивает то, что VAR2[] приведенные выше термины следуют шаблону суммы префикса.

Это означает, что один из возможных методов решения может быть задан следующим образом:

VAR1[i] = 1+prefix_sum(VAR2[2K + (i-2)])   (for i > 0) notes:(1) (2)
VAR1[i] = 1                                (for i = 0)

Теперь префиксная сумма может быть выполнена параллельно (это не совсем независимая операция, но она может быть распараллелена. Я не хочу спорить здесь слишком много о терминологии или чистоте. Я предлагаю один из возможных способов распараллеливания для вашей заявленной проблемы, а не единственный способ сделать это.) Чтобы сделать сумму префикса параллельно на GPU, я бы использовал библиотеку, как CUB или Thrust. Или вы можете написать свой, хотя я бы не советовал.

Заметки:

1. использование -1 или -2 в качестве смещения i для суммы префикса может быть продиктовано вашим использованием inclusive или же exclusive операция сканирования или префикса суммы.

2. VAR2 должен быть определен над соответствующей областью, чтобы сделать это разумным. Однако это требование подразумевается в вашей постановке задачи.

Вот тривиальный проработанный пример. В этом случае, так как VAR2 термин индексации 2K+(I-1) просто представляет фиксированное смещение I (2K-1), мы просто используем смещение 0 для демонстрационных целей, поэтому VAR2 это просто массив в том же домене, что и VAR1, И я определяю VAR2 просто быть массивом всего 1, в демонстрационных целях. Параллельное вычисление GPU происходит в VAR1 вектор, вычисление эквивалента процессора просто вычисляется на лету в cpu переменная для целей проверки:

$ cat t1056.cu
#include <thrust/scan.h>
#include <thrust/device_vector.h>
#include <thrust/host_vector.h>
#include <thrust/transform.h>
#include <iostream>

const int dsize = 1000;
using namespace thrust::placeholders;
int main(){

  thrust::device_vector<int> VAR2(dsize, 1);  // initialize VAR2 array to all 1's
  thrust::device_vector<int> VAR1(dsize);
  thrust::exclusive_scan(VAR2.begin(), VAR2.end(), VAR1.begin(), 0); // put prefix sum of VAR2 into VAR1
  thrust::transform(VAR1.begin(), VAR1.end(), VAR1.begin(),  _1 += 1);   // add 1 to every term
  int cpu = 1;
  for (int i = 1; i < dsize; i++){
    int gpu = VAR1[i];
    cpu += VAR2[i];
    if (cpu != gpu) {std::cout << "mismatch at: " << i << " was: " << gpu << " should be: " << cpu << std::endl; return 1;}
    }
  std::cout << "Success!" << std::endl;
  return 0;
}

$ nvcc -o t1056 t1056.cu
$ ./t1056
Success!
$

Для получения дополнительной ссылки, в частности, на использование операций сканирования для решения задач линейного повторения, см. Статью Blelloch здесь, раздел 1.4.