CUDA 并行化答案

【问题标题】：CUDA parallelizationCUDA 并行化
【发布时间】：2012-04-26 07:22:20
【问题描述】：

我在使用 CUDA 对数字数组进行并行化时遇到问题。

因此，例如，如果我们有一个包含数字（1、2、3、4、5）的数组 M

如果我要删除数组中的数字 2 并将所有内容向左移动，结果数组将是 (1 , 3 , 4 , 5 , 5 )

其中 M[1] = M[2]，M[2] = M[3]，M[3] = M[4]

我的问题是我们如何在 cuda 中并行执行此操作？因为当我们并行这个可能存在竞争条件，其中数字 2 (M[1]) 可能不是第一个首先行动，如果 M[2] 是第一个移动的，则结果数组将变为 (1、4、4、5、5)。有什么方法可以处理这个吗？我对 cuda 还很陌生，所以我不知道该怎么办...

我目前的代码如下：

__global__ void gpu_shiftSeam(int *MCEnergyMat, int *seam, int width, int height, int currRow)
{
    int i = blockIdx.x * blockDim.x + threadIdx.x;
    int j = blockIdx.y * blockDim.y + threadIdx.y;

    int index = i+width*j;

    if(i < width && j <height)
    {
        //shift values of -1 to the side of the image
        if(MCEnergyMat[i+width*j] == -1)
        {
            if(i+1 != width)
                    MCEnergyMat[index] = MCEnergyMat[index+1];
        }
        if(seam[j] < i)
        {
            if(i+1 != width)
                MCEnergyMat[index] = MCEnergyMat[index+1];
        }
    }
}

seam[i] 包含我想在数组中删除的索引。而MCEnergyMat 只是从二维数组转换而来的一维数组......但是，我的代码不起作用......我相信竞争条件是问题所在。

谢谢！

【问题讨论】：

流压缩是 GPU 上已解决的问题。您可以使用许多强大的现成 CUDA 实现，包括几年来随 CUDA 工具包一起提供的 thrust 实现。你为什么不只使用其中之一？

标签： cuda

【解决方案1】：

正如 talonmies 在他的评论中指出的那样，这种事情被称为“流压缩”。使用 Thrust 的方法如下：

#include <thrust/device_vector.h>
#include <thrust/host_vector.h>
#include <thrust/remove.h>
#include <iostream>

int main()
{
  int data[5] = {1,2,3,4,5};
  thrust::device_vector<int> d_vec(data, data + 5);

  // new_end points to the end of the sequence after 2 has been thrown out
  thrust::device_vector<int>::iterator new_end = 
    thrust::remove(d_vec.begin(), d_vec.end(), 2);

  // erase everything after the new end
  d_vec.erase(new_end, d_vec.end());

  // prove that it worked
  thrust::host_vector<int> h_vec = d_vec;

  std::cout << "result: ";
  thrust::copy(h_vec.begin(), h_vec.end(), std::ostream_iterator<int>(std::cout, " "));
  std::cout << std::endl;

  return 0;
}

结果如下：

$ nvcc test.cu -run result: 1 3 4 5

【讨论】：