c++ - 推力/cuda reduce_by_key 错误？

Question

我遇到了 thrust 库的 reduce_by_key 函数的问题。对我来说这看起来像是一个错误，但我想在报告之前确定一下。

首先，我的设置:CUDA 7.0、Windows 8、NIVIDA GeForce 820m。整个过程是使用 visual studio 2010 和 nvcc 在 Release模式下编译的，64 位。

现在，说明问题的练习。

我的设备上生成了一个名为 devData 的随机数 vector 。我将一个名为 devIndices 的索引 vector 制成表格，其大小相同，定义如下:

• devIndices = [0, 0, 0, 0, 1, 1, 1, 1, ... K-1, K-1, K-1, K-1]
• devData = [ 1, 4, 5, 7, 5, 8, 9, 6, ... 7, 8, 9, 6]

因此，在本例中，devIndices 中的每个值都会重复 mod = 4 次。

然后，我只想使用 devIndices 减少_by_key devData 以获得以下减少的 vector :

• devIndices = [0, 1, ..., K-1]
• 开发数据 = [17, 28,..., 30]

(如果我的算术是正确的 :) )

现在，我确定 devIndices 的元素总和应该等于由以下关系给出的值 T:

• T = [(K-1) * K/2](例如:[0 1 2 3] -> 6 = (K-1)*K/2 = 3 * 4/2)

我尝试在我的机器上执行此操作，它适用于少量元素，但无法用于大量元素。 (100,000 次失败...)

下面是我用来如上所述操作我的两个 vector 并在最后输出 devIndices 之和的代码。您可以使用基本上设置元素数量的参数 k。

#include <cuda.h>
#include <thrust/random.h>
#include <thrust/device_vector.h>
#include <thrust/sort.h>
#include <thrust/iterator/counting_iterator.h>
#include <fstream>
typedef typename thrust::device_vector<int>     tDevVecInt;
typedef typename thrust::device_vector<float>   tDevVecFlt;

struct rando : public thrust::unary_function<unsigned int, float>
{
    unsigned int mainSeed;
    rando(unsigned int _mainSeed):mainSeed(_mainSeed) {}
    __host__ __device__ float operator()(unsigned int x) 
    {
        unsigned int seed = x * mainSeed;
        thrust::random::taus88 mac(seed);
        thrust::uniform_real_distribution<float> dist(0,1);
        return dist(mac);
    }
};

struct modSim : public thrust::unary_function<int, int>  
{
    int sz;
    modSim(int in)
    {
        this->sz = in;
    }
    __host__ __device__ int operator()(const int &x) 
    {
        return x/sz;
    }
};

int main() 
{
    int mod = 10;
    int k = 10000;
    int szData = k*mod;

    tDevVecFlt devData(szData, 0.);
    tDevVecInt devIndices(szData, 0.);

    thrust::transform(thrust::make_counting_iterator(0), thrust::make_counting_iterator(0) + szData, devData.begin(), rando(123456789));    
    thrust::tabulate(devIndices.begin(), devIndices.end(), modSim(mod)); 
    thrust::reduce_by_key(devIndices.begin(), devIndices.end(), devData.begin(), devIndices.begin(), devData.begin());
    std::cout << thrust::reduce(devIndices.begin(), devIndices.begin()+ k, 0) << std::endl;
    return 0;
}

最糟糕的是:当我多次运行同一段代码时，我得到了不同的结果! random vector跟这个没关系(是seed...顺便查了一下)。

那么现在的问题部分:

• 我哪里错了吗？ Reduce_by_key 对我来说似乎是正确的工具
• 有人重现这种不可再现性吗？
• 如果这确实是一个错误，通常的报告方式是什么？

Answer 1

Am I wrong somewhere?

documentation对于 thrust::reduce_by_key 状态:

Precondition The input ranges shall not overlap either output range.

你在你的代码中打破了这个先决条件:

thrust::reduce_by_key(devIndices.begin(), devIndices.end(), devData.begin(), devIndices.begin(), devData.begin());

因此您的代码已损坏，并不代表任何展示推力错误的情况。 thrust::reduce_by_key 不是可以就地完成的推力操作。