为什么我不能用numba（cuda python）获得1D数组的正确和？

Question

我试着用cuda python和numba。代码是计算1D的和数组，但我不知道如何得到一个值结果而不是
三个价值观。python3.5和Nuba+CUDA8.0

import os,sys,time
import pandas as pd
import numpy as np
from numba import cuda, float32

os.environ['NUMBAPRO_NVVM']=r'D:\NVIDIA GPU Computing Toolkit\CUDA\v8.0\nvvm\bin\nvvm64_31_0.dll'
os.environ['NUMBAPRO_LIBDEVICE']=r'D:\NVIDIA GPU Computing Toolkit\CUDA\v8.0\nvvm\libdevice'

bpg = (1,1) 
tpb = (1,3)

@cuda.jit
def calcu_sum(D,T):
    ty = cuda.threadIdx.y
    bh = cuda.blockDim.y
    index_i = ty
    L = len(D)
    su = 0
    while index_i<L:
        su +=D[index_i]
        index_i +=bh
    print('su:',su)
    T[0,0]=su
    print('T:',T[0,0])


D = np.array([ 0.42487645,0.41607881,0.42027071,0.43751907,0.43512794,0.43656972,
               0.43940639,0.43864551,0.43447691,0.43120232], dtype=np.float32)
T = np.empty([1,1])
print('D: ',D)

stream = cuda.stream()
with stream.auto_synchronize():
    dD = cuda.to_device(D, stream)
    dT= cuda.to_device(TE, stream)
    calcu_sum[bpg, tpb, stream](dD,dT)

The output is:

D:  [ 0.42487645  0.41607881  0.42027071  0.43751907  0.43512794  0.43656972
  0.43940639  0.43864551  0.43447691  0.43120232]
su:  1.733004
su:  1.289852
su:  1.291317
T: 1.733004
T: 1.289852
T: 1.291317

为什么我不能得到输出“4.31417383”而不是“1.733004 1.289852”
1.291317” ? 1.733004+1.289852+1.291317=4.314173.
我是新来的麻木，阅读麻木文件，但不知道怎么做。
有人能给点建议吗？

Answer 1

你没有得到你所期望的总数是因为你没有写信
代码生成那个总和。
基本的CUDA编程模型（无论您是使用cudac、Fortran还是Python作为
你的语言）是你编写内核代码，由每个线程执行。
您已经为每个线程编写了代码来读取和求和输入数组的一部分。
您还没有为这些线程编写任何代码来共享和求和它们的
个别部分和变成最终和。
有一个非常好的算法来实现这一点——它是
称为平行还原。您可以在中找到该算法的介绍
在CUDA工具包的每个版本的示例中提供的PDF，或
下载相关演示文稿
这里.
你也可以阅读一个更现代版本的算法，它使用了更新的
CUDA的特点（warp-shuffle指令和原子事务）
[这里](https://devblogs.nvidia.com/parallelforall/faster-parallel-reductions-
开普勒/）。
在学习了约简算法之后，您将需要调整
将标准cudac内核代码转换成Numba Python内核方言。光秃秃的
最低限度，类似这样：

tpb = (1,3)

@cuda.jit
def calcu_sum(D,T):

    ty = cuda.threadIdx.y
    bh = cuda.blockDim.y
    index_i = ty
    sbuf = cuda.shared.array(tpb, float32)

    L = len(D)
    su = 0
    while index_i < L:
        su += D[index_i]
        index_i +=bh

    print('su:',su)

    sbuf[0,ty] = su
    cuda.syncthreads()

    if ty == 0:
        T[0,0] = 0
        for i in range(0, bh):
            T[0,0] += sbuf[0,i]
        print('T:',T[0,0])

可能会做你想做的事，尽管离成功还有很长的路要走
最佳的并行共享内存减少，当您阅读
我提供链接的材料。