Замена c для цикла на cuda

Question

Каков наилучший способ сделать это в CUDA?

...
for(int i=0;i<size;++i)                                                                             
  for(int j=i+1;j<size ;++j)                                                                           
    temp_norm+=exp((train[i]-train[j])/tau);   

Это будет эквивалентно?

...
int i = threadIdx.x + blockIdx.x * blockDim.x;
int j = threadIdx.y + blockIdx.y * blockDim.y;

if (i>=size || j>=size) return;

if(j>i)
  temp_norm+=exp((train[i]-train[j])/tau);

Любая помощь будет высоко ценится!

Answer 1

Как лучше реализовать, зависит от того, насколько велик size. Но при условии, что он довольно большой, например, 1000 или больше ...

Чтобы сделать это так, как вы предлагаете, вам нужно использовать atomicAdd (), который может быть дорогим, если слишком много потоков атомарно добавляются к одному и тому же адресу. Возможно, лучше использовать параллельное сокращение.

Проверьте "уменьшение" семпла в NVIDIA CUDA SDK.

YMMV со следующим, поскольку он не проверен, и я не знаю ваш размер данных, но что-то вроде этого должно работать. Воспользуйтесь ядром "extract6" из этого примера, но добавьте свои вычисления в первый цикл while. Заменить инициализацию i и gridSize на

unsigned int i = blockIdx.x*blockSize + threadIdx.x;
unsigned int gridSize = blockSize * gridDim.x;

Заменить петлю while (i < n) на

while (i < size)
{
  for (unsigned int j = i+1; j<size; ++j)
      mySum += exp((train[j]-train[i])/tau);   
  i += gridSize;
}

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