在设备阵列上按键减少

Question

我正在使用 reduce_by_key 查找具有相同第一个值的 int2 类型数组中的元素数。

例如数组：<1,2> <1,3> <1,4> <2,5> <2,7> 所以不行。以 1 作为第一个元素的元素是 3，以 2 为第一个元素的元素是 2。

代码：

struct compare_int2 : public thrust::binary_function<int2, int2, bool> {
__host__ __device__ bool operator()(const int2 &a,const int2 &b) const{
return (a.x == b.x);}
};

compare_int2 cmp;
int main()
{
        int n,i;
        scanf("%d",&n);

        int2 *h_list = (int2 *) malloc(sizeof(int2)*n);
        int *h_ones = (int *) malloc(sizeof(int)*n);

        int2 *d_list,*C;
        int  *d_ones,*D;
        cudaMalloc((void **)&d_list,sizeof(int2)*n);
        cudaMalloc((void **)&d_ones,sizeof(int)*n);
        cudaMalloc((void **)&C,sizeof(int2)*n);
        cudaMalloc((void **)&D,sizeof(int)*n);

        for(i=0;i<n;i++)
        {
                int2 p;
                printf("Value ? ");
                scanf("%d %d",&p.x,&p.y);
                h_list[i] = p;
                h_ones[i] = 1;
        }

        cudaMemcpy(d_list,h_list,sizeof(int2)*n,cudaMemcpyHostToDevice);
        cudaMemcpy(d_ones,h_ones,sizeof(int)*n,cudaMemcpyHostToDevice);
        thrust::reduce_by_key(d_list, d_list+n, d_ones, C, D,cmp);
        return 0;
}

以上代码显示了 Segmentation Fault 。我运行上面的代码使用 gdb，它报告了这个位置的段错误。

thrust::system::detail::internal::scalar::reduce_by_key > (keys_first=0x1304740000,keys_last=0x1304740010,values_first=0x1304740200,keys_output=0x1304740400, values_output=0x1304740600,binary_pred=...,binary_op=...)

at /usr/local/cuda-6.5/bin/../targets/x86_64-linux/include/thrust/system/detail/internal/scalar/reduce_by_key.h:61 61
InputKeyType temp_key = *keys_first

如何在设备阵列上使用reduce_by_key？

Answer 1

Thrust 将普通指针解释为指向主机上的数据：

    thrust::reduce_by_key(d_list, d_list+n, d_ones, C, D,cmp);

因此 thrust 将为上述算法调用 主机路径 ，并且当它试图取消引用主机代码中的那些指针时会出现段错误。这包含在 thrust getting started guide:

You may wonder what happens when a "raw" pointer is used as an argument to a Thrust function. Like the STL, Thrust permits this usage and it will dispatch the host path of the algorithm. If the pointer in question is in fact a pointer to device memory then you'll need to wrap it with thrust::device_ptr before calling the function.

Thrust 有多种机制（例如 device_ptr、device_vector 和执行策略）来向算法识别数据是设备驻留的并且应该使用设备路径。

对现有代码最简单的修改可能是使用 device_ptr:

#include <thrust/device_ptr.h>
...
thrust::device_ptr<int2> dlistptr(d_list);
thrust::device_ptr<int>  donesptr(d_ones);
thrust::device_ptr<int2> Cptr(C);
thrust::device_ptr<int>  Dptr(D);
thrust::reduce_by_key(dlistptr, dlistptr+n, donesptr, Cptr, Dptr,cmp);

上述问题类似于another issue you asked about。

在设备阵列上按键减少

Reduce by key on device array

cuda

parallel-processing

thrust