OpenCL中本地和全局內存區別有什麼區別？

我試圖用PyOpenCL作出一個減少總和，類似於這個例子：https://dournac.org/info/gpu_sum_reduction。我試圖對所有值爲1的矢量求和。第一個元素的結果應該是16384。但是，似乎只有一些要點正在收集。是否需要本地索引？是否有任何競爭條件（當我運行兩次結果是不一樣的）？下面的代碼有什麼問題？OpenCL中本地和全局內存區別有什麼區別？

import numpy as np 
import pyopencl as cl 

def readKernel(kernelFile): 
    with open(kernelFile, 'r') as f: 
     data=f.read() 
    return data 

a_np = np.random.rand(128*128).astype(np.float32) 
a_np=a_np.reshape((128,128)) 
print(a_np.shape) 

device = cl.get_platforms()[0].get_devices(cl.device_type.GPU)[0] 
print(device) 
ctx=cl.Context(devices=[device]) 
#ctx = cl.create_some_context() #ask which context to use 
queue = cl.CommandQueue(ctx) 
mf = cl.mem_flags 

a_g = cl.Buffer(ctx, mf.READ_WRITE | mf.COPY_HOST_PTR, hostbuf=a_np) 

prg = cl.Program(ctx,readKernel("kernel2.cl")).build() 

prg.test(queue, a_np.shape, None, a_g) 

cl.enqueue_copy(queue, a_np, a_g).wait() 
np.savetxt("teste2.txt",a_np,fmt="%i")

內核是：

__kernel void test(__global float *count){ 
    int id = get_global_id(0)+get_global_id(1)*get_global_size(0); 
    int nelements = get_global_size(0)*get_global_size(1); 

    count[id] = 1; 
    barrier(CLK_GLOBAL_MEM_FENCE); 

    for (int stride = nelements/2; stride>0; stride = stride/2){ 
     barrier(CLK_GLOBAL_MEM_FENCE); //wait everyone update 
     if (id < stride){ 
      int s1 = count[id]; 
      int s2 = count[id+stride]; 
      count[id] = s1+s2; 
     } 
    } 
    barrier(CLK_GLOBAL_MEM_FENCE); //wait everyone update 
}

來源

2017-09-22 Rubens Sautter

的問題是，你的內核中實現一個工作組內做的減少和有隱含schedulled許多工作組。

取決於GPU，每個工作組的最大工作項目數量不同。對於Nvidia 1024，AMD和Intel 256（英特爾在更早的GPU中有512）。

讓我們假設在這個例子中，GPU上每個工作組的最大工作項是256.在這種情況下，最大二維worgroup大小可以是16x16，所以如果使用矩陣的大小，內核將返回正確的結果。在調度內核時，使用原始大小爲128x128而不指定本地大小的實現計算出，對於您和您將獲得全局大小128x128和本地大小（非常可能）16x16，這意味着正在調度8個worgroup。在當前的內核中，每個工作組都從不同的id開始計算，但索引被減少到0，所以你有競爭條件，因此每次運行都會產生不同的結果。

您有2個選項來解決這個問題：

重寫內核一個工作組中的所有計算以及與全球，當地的大小安排它：（16×16），（16,16）或任何你最大的工作每個工作組設備上的項目有
使用全局，本地大小：（128x128），（16x16），並且每個工作組將計算其結果，然後在cpu端將必須總結每個工作組以獲得最終結果。

對於128x128，第一個選項將是首選，因爲它應該執行得更快，應該更直接地實現。

來源

2017-09-23 07:07:30 doqtor

OpenCL中本地和全局內存區別有什麼區別？

回答

相關問題