cuda计算的结果时时变化

Question

对不起我的英语。我有一个 cuda 内核，它不时返回不同的结果值。该内核计算系列和。我的内核由 4 个代码部分组成。让我解释一下这个内核是如何工作的。第一部分在线程之间分配迭代（I took it as source）。第二个代码部分显示了每个线程如何计算半数。在第二部分之后我们必须放置 __syncthreads() 因为在第二部分之后我们开始使用共享内存。在第三部分中，我得到块中所有线程的总和，并将其放入 threadIdx.x 等于 0(I took it as source @ page 22) 的线程。在第四部分中，我得到所有线程块的总和并将其放入 dSum[0]

我是否正确放置了 __syncthreads()？哪里有错误？为什么在 64 个块和 768 个线程上给出错误的结果，而在 768 个块和 64 个线程上给出正确的结果？

__global__ void sumSeries(double* dSum,int totalThreadNumber){  
    volatile  __shared__ double data[768];  
    int tid=threadIdx.x+blockIdx.x*blockDim.x;
    int myend;
    double var;
    //part_1 get tid's start iteration value and end iteration value.
    int mystart = (INT_MAX / totalThreadNumber) * tid;
    if (INT_MAX %  totalThreadNumber > tid)
    {
        mystart += tid;
        myend = mystart + (INT_MAX /  totalThreadNumber) + 1;
    }
    else
    {
        mystart += INT_MAX %  totalThreadNumber;
        myend = mystart + (INT_MAX /  totalThreadNumber);
    }
    //part_2 get halfsum
    data[threadIdx.x]=0;
    for (int i = mystart ; i < myend ; ++i){
            var=i;
            data[threadIdx.x] += (var*var+var+1)/(var*var*var+var*var+var+1);

    }   
    __syncthreads();

    //part_3 sum all results in every block
    for (int s=blockDim.x/2; s>32; s>>=1)
    {
        if (threadIdx.x < s)
            data[threadIdx.x] += data[threadIdx.x + s];
        __syncthreads();
    }
    if (threadIdx.x < 32)
    {
        data[threadIdx.x] += data[threadIdx.x + 32];
        data[threadIdx.x] += data[threadIdx.x + 16];
        data[threadIdx.x] += data[threadIdx.x + 8];
        data[threadIdx.x] += data[threadIdx.x + 4];
        data[threadIdx.x] += data[threadIdx.x + 2];
        data[threadIdx.x] += data[threadIdx.x + 1];
    }

    if (threadIdx.x==0)
    {
        dSum[blockIdx.x]=data[0];
    }
    __syncthreads();
    //part_4
    if (tid==0)
        for (int t=1;t<8;++t)
            dSum[0]=dSum[0]+dSum[t];
}

Answer 1

所以你的总和是系列结束

(n^2+n+1)/(n^3+n^2+n+1) = (n^3-1)/(n^4-1)

这与调和级数具有相同的收敛性

1/n

即没有，它非常非常缓慢地向无穷大发散。从 1 到 N 的和具有 log(N) 和 1-log(2)+log(N+1) 之间的值。

这些级数的任何有限求和的结果对于求和的顺序是非常明智的。从 1 到 N 向前求和并减少会抑制 1==1+1/n 的所有项，对于浮点数而言，这发生在一个相当小的数字上。从某个 N 倒数到 1 将首先累积小数并保留它们的累积贡献。

所以根据部分和的到达顺序，特别是当包含1的和进来时，总和会出现明显的差异。

---

这两项在

中都是单调递减的

f(x) = (x^2+x+1)/(x^3+x^2+x+1) = 0.5/(x+1)+0.5*(x+1)/(x^2+1)

该函数的反导数是

F(n) = 0.5*ln(x+1)+0.25*ln(x^2+1)+0.5*arctan(x)

这样

f(n+1) <= F(n+1)-F(n) <= f(n) <= F(n)-F(n-1)

总结结果

F(N+1)-F(m) <= sum(n=m to N) f(n) <= F(N)-F(m-1)

为此，必须在所有三个术语中添加总和的初始部分。

所以设置 m=1000, computeS=sum(n=0 to 999) f(n)`，然后

S+F(2^32  )-F(1000) = 23.459829390459243
S+F(2^32-1)-F( 999) = 23.460829890558995

是从 0 到 2^32-1 求和的上限和下限，远离任何数值结果。