Блог

Главная - Вопросы по программированию - плюс равный ( =) оператор в pycuda

плюс равный ( =) оператор в pycuda

#python #parallel-processing #cuda #gpu #pycuda

#python #параллельная обработка #cuda #графический процессор #pycuda

Вопрос:

Я хотел бы реализовать вариант свертки в pycuda.

Для простоты я покажу прямоугольное ядро интерполяции. Стандартная свертка может быть применена следующим образом:

 import pycuda.autoinit
import pycuda.driver as drv
import numpy as np
from pycuda.compiler import SourceModule

mod = SourceModule("""
#include <stdio.h>
__global__ void func(float *dest, float *a)
{
  const int img_size = 64;
  const int kernel_size = 3;
  const int kernel_size_half = kernel_size/2;
  const int tx = blockIdx.x * blockDim.x   threadIdx.x;
  const int ty = blockIdx.y * blockDim.y   threadIdx.y;
  int tx_kernel;

  tx_kernel = tx - kernel_size_half;
  for (int idx=-kernel_size_half; idx <= kernel_size_half; idx  ) 
  {
      tx_kernel = tx   idx ;
      if ((tx_kernel < 0) || (tx_kernel > img_size-1))
        continue;
      dest[ty * img_size   tx]  =  a[ty * img_size   tx_kernel] / ((float) kernel_size);
  }
}
""")

Вместо того, чтобы вычислять текущую позицию по соседям, я хотел бы сделать наоборот,
добавить значение текущего пикселя к соседям.

То есть:

чтобы изменить строку:

 dest[ty * img_size   tx]  =  a[ty * img_size   tx_kernel] / ((float) kernel_size);

Для:

 dest[ty * img_size   tx_kernel]  =  a[ty * img_size   tx] / ((float) kernel_size);

Однако первый работает нормально, а второй — нет, он завершается сбоем при обновлении соседей.
Есть ли способ обойти его?

Примечание: я упростил вопрос, чтобы сосредоточиться на том, что мне нужно, общая проблема заключается в использовании другого ядра свертки для каждого пикселя вместо того, которое я задал в вопросе.

Комментарии:

1. «Вместо того, чтобы вычислять текущую позицию по соседям, я хотел бы сделать наоборот», как правило, это более сложный подход, потому что теперь у вас есть возможность одновременного обновления нескольких потоков в одном и том же местоположении. Для этого может потребоваться атомизация или какой-либо другой метод для сортировки, что обычно приводит к более медленной реализации. Тем не менее, если вы хотите сделать это по какой-то причине, я бы начал с обновления соседей с помощью atomics.

2. вместо: dest[ty * img_size tx_kernel] = a[ty * img_size tx] / ((float) kernel_size); сделайте это: atomicAdd(amp;(dest[ty * img_size tx_kernel]), a[ty * img_size tx] / ((float) kernel_size));

Ответ №1:

чтобы изменить строку:

 dest[ty * img_size   tx]  =  a[ty * img_size   tx_kernel] / ((float) kernel_size);

Для:

 dest[ty * img_size   tx_kernel]  =  a[ty * img_size   tx] / ((float) kernel_size);

Однако первый работает нормально, а второй — нет, он завершается сбоем при обновлении соседей. Есть ли способ обойти его?

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

 atomicAdd(amp;(dest[ty * img_size   tx_kernel]), a[ty * img_size   tx] / ((float) kernel_size));