CUDA: Как можно извлечь определенные точки изображения в непрерывный вектор?

Question

Я хочу написать функцию CUDA, которая извлекает точки изображения, которые удовлетворяют определенному условию, а затем помещает их в непрерывный блок памяти на устройстве.

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

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

Если у вас есть какие-либо предложения о том, как я могу взять набор точек и перенести их в новое местоположение на устройстве параллельно (!), Я открыт для идей. Спасибо за ваше время.

Answer 1

Вот общий способ сделать это:

• Ядро 1: Извлечь точки изображения и записать их в массив.Результат в массиве не является непрерывным.
• Соберите точки в несмежном массиве в непрерывный массив.
• Ядро 2: Работа с точками изображения.

Ядро 1 могло записать точки изображения в массив с пробелами, поскольку вы не можете предсказать, сколько точек изображения получится.Таким образом, вам нужно собрать записанные изображения вместе, прежде чем запускать на нем ядро ​​2.Сбор довольно прост в использовании, если вы используете библиотеку типа Thrust .Например, его функция remove_if может использоваться для удаления точек, помеченных как недействительные или пустые.

Answer 2

Вы также можете попытаться добавить результаты непосредственно в выходной вектор, используя атомарные функции, т. Е .:

__global__ void kernel( dataType *inputImage, dataType *a, int *sizeof_A)
{
    // map from threadIdx/BlockIdx to pixel position
    int x = threadIdx.x + blockIdx.x * blockDim.x;
    int y = threadIdx.y + blockIdx.y * blockDim.y;
    int offset = x + y * blockDim.x * gridDim.x;

    if (inputImage[offset] == /* your condition */ ) {
        int arrayLastPosition = atomicAdd(sizeof_A, 1);
        a[arrayLastPosition] = /* your mark */;
    }
}

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