使用PyOpenCL將結構與指針成員傳遞給OpenCL內核

假設我有一個內核來計算兩個數組的元素總和。而不是將A，B和C三個參數，我讓他們結構成員如下：使用PyOpenCL將結構與指針成員傳遞給OpenCL內核

typedef struct 
{ 
    __global uint *a; 
    __global uint *b; 
    __global uint *c; 
} SumParameters; 

__kernel void compute_sum(__global SumParameters *params) 
{ 
    uint id = get_global_id(0); 
    params->c[id] = params->a[id] + params->b[id]; 
    return; 
}

有結構上的信息，如果您PyOpenCL的RTFM [1]，和其他人也解決了這個問題[ 2] [3] [4]。但是，我找到的OpenCL結構示例都沒有指針作爲成員。

具體來說，我很擔心主機/設備地址空間是否匹配，以及主機/設備指針大小是否匹配。有人知道答案嗎？

[1] http://documen.tician.de/pyopencl/howto.html#how-to-use-struct-types-with-pyopencl

[2] Struct Alignment with PyOpenCL

[3] http://enja.org/2011/03/30/adventures-in-opencl-part-3-constant-memory-structs/

[4] http://acooke.org/cute/Somesimple0.html

來源

2013-07-14 picomancer

不，地址空間不匹配。對於基本類型（float，int，...），您有對齊要求（標準的第6.1.5節），您必須使用OpenCL實現的cl_type名稱（在C編程時，pyopencl在引擎蓋下執行作業I說吧）。

對於指針來說，它更簡單，因爲這種不匹配。標準v 1第6.9節的開頭部分。2（這是1.1版本第6.8節）指出：

參數傳遞給內核是指針必須用__global，__constant或__local預選賽宣告程序中聲明的函數。

而且在點p：

參數傳遞給內核被宣佈爲一個結構或工會不允許的OpenCL對象爲結構或聯合的元素傳遞功能。

還要注意點d：

可變長度數組和結構具有柔性（或未施膠）陣列不被支持。

所以，沒有辦法讓你內核運行在你的問題中描述，這就是爲什麼你一直沒能找到的OpenCL結構的一些例子有一個指針作爲成員。
我仍然可以提出一個解決方法，它利用內核在JIT中編譯的事實。它仍然要求您正確打包數據，並注意對齊，並最終在執行程序期間大小不會改變。老實說，我會選擇一個以3個緩衝區爲參數的內核，但無論如何，它就是這樣。

的想法是使用預處理器選項-D在下面的例子中的Python：

內核：

typedef struct { 
    uint a[SIZE]; 
    uint b[SIZE]; 
    uint c[SIZE]; 
} SumParameters; 

kernel void foo(global SumParameters *params){ 
    int idx = get_global_id(0); 
    params->c[idx] = params->a[idx] + params->b[idx]; 
}

主機代碼：

import numpy as np 
import pyopencl as cl 

def bar(): 
    mf = cl.mem_flags 
    ctx = cl.create_some_context() 
    queue = cl.CommandQueue(self.ctx) 
    prog_f = open('kernels.cl', 'r') 
    #a = (1, 2, 3), b = (4, 5, 6)   
    ary = np.array([(1, 2, 3), (4, 5, 6), (0, 0, 0)], dtype='uint32, uint32, uint32') 
    cl_ary = cl.Buffer(ctx, mf.READ_WRITE | mf.COPY_HOST_PTR, hostbuf=ary) 
    #Here should compute the size, but hardcoded for the example 
    size = 3 
    #The important part follows using -D option 
    prog = cl.Program(ctx, prog_f.read()).build(options="-D SIZE={0}".format(size))  
    prog.foo(queue, (size,), None, cl_ary) 
    result = np.zeros_like(ary) 
    cl.enqueue_copy(queue, result, cl_ary).wait() 
    print result

而結果：

[(1L, 2L, 3L) (4L, 5L, 6L) (5L, 7L, 9L)]

來源

2013-07-15 13:43:27 CaptainObvious

我不知道答案，我自己的問題，但有是3種解決方法，我可以拿出我的頭頂。我認爲解決方法3是最佳選擇。

解決方法1：我們在這裏只有3個參數，所以我們可以製作a，b和c內核參數。但我已經讀過，可以傳遞給內核的參數數量是有限的，我個人喜歡重構任何需要超過3-4個參數才能使用結構的函數（或者，在Python中，元組或關鍵字參數）。所以這個解決方案使得代碼難以閱讀，並且不能縮放。

解決方法2：將所有內容轉儲到一個巨型陣列中。那麼內核是這樣的：

typedef struct 
{ 
    uint ai; 
    uint bi; 
    uint ci; 
} SumParameters; 

__kernel void compute_sum(__global SumParameters *params, uint *data) 
{ 
    uint id = get_global_id(0); 
    data[params->ci + id] = data[params->ai + id] + data[params->bi + id]; 
    return; 
}

換句話說，而不是使用指針，使用偏移到一個單一的陣列。這看起來非常像實現我自己的內存模型的開始，它感覺像是重新發明了一個存在於PyOpenCL或OpenCL或兩者中的輪子。

解決方法3：製作setter內核。就像這樣：

__kernel void set_a(__global SumParameters *params, __global uint *a) 
{ 
    params->a = a; 
    return; 
}

，並同上，用於set_b，set_c。然後用worksize 1執行這些內核來設置數據結構。你仍然需要知道爲params分配多大的塊，但是如果它太大，沒有什麼不好的事情會發生（除了一點點浪費的內存），所以我只是假設指針是64位。

這種解決方法的性能可能很糟糕（我想象一個內核調用會有巨大的開銷），但幸運的是，對於我的應用程序來說不應該太重要（我的內核一次將運行數秒，它不是圖形必須以30-60 fps的速度運行，所以我想象一下，額外的內核調用設置參數所花費的時間最終將成爲我工作負載的一小部分，不管每個內核調用開銷多高）。

來源

2013-07-14 02:04:38 picomancer

使用PyOpenCL將結構與指針成員傳遞給OpenCL內核

回答

相關問題