¿Por qué no puedo obtener la sum correcta de la matriz 1D con numba (cuda python)?

Intento usar cuda python con numba. El código es calcular la sum de una matriz 1D de la siguiente manera, pero no sé cómo obtener un resultado de valor en lugar de tres valores.

python3.5 con numba + 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) 

La salida es:

 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 

¿Por qué no puedo obtener la salida “4.31417383” en lugar de “1.733004 1.289852 1.291317”? 1.733004 + 1.289852 + 1.291317 = 4.314173.

Soy nuevo en numba, lea la documentación de numba, pero no sé cómo hacerlo. ¿Alguien puede dar consejos?

La razón por la que no obtiene la sum que espera es porque no ha escrito código para producir esa sum.

El modelo de progtwigción básico de CUDA (ya sea que uses CUDA C, Fortran o Python como tu idioma) es que escribes el código del kernel que se ejecuta en cada hilo. Ha escrito código para cada hilo para leer y sumr parte de la matriz de entrada. No ha escrito ningún código para que esos hilos compartan y sumen sus sums parciales individuales en una sum final.

Hay un algoritmo extremadamente bien descrito para hacer esto: se llama reducción paralela. Puede encontrar una introducción al algoritmo en un PDF que se incluye en los ejemplos de cada versión del kit de herramientas CUDA, o descargar una presentación al respecto aquí . También puede leer una versión más moderna del algoritmo que usa características más nuevas de CUDA (instrucciones de mezcla aleatoria y transacciones atómicas) aquí .

Una vez que haya estudiado el algoritmo de reducción, deberá adaptar el código estándar del kernel CUDA C al dialecto del kernel Numba Python. Como mínimo, algo como esto:

 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]) 

probablemente hará lo que quiera, aunque aún está muy lejos de una reducción de memoria compartida paralela óptima, como verá cuando lea el material al que proporcioné los enlaces.