2011-08-25 71 views
10

我與設在這裏的教程沿着以下:http://opencl.codeplex.com/wikipage?title=OpenCL%20Tutorials%20-%201OpenCL - 是否可以在內核中調用另一個函數?

他們列出的內核是這樣的,它計算在輸出變量兩個數字,並將其存儲的總和:

__kernel void vector_add_gpu (__global const float* src_a, 
        __global const float* src_b, 
        __global float* res, 
      const int num) 
{ 
    /* get_global_id(0) returns the ID of the thread in execution. 
    As many threads are launched at the same time, executing the same kernel, 
    each one will receive a different ID, and consequently perform a different computation.*/ 
    const int idx = get_global_id(0); 

    /* Now each work-item asks itself: "is my ID inside the vector's range?" 
    If the answer is YES, the work-item performs the corresponding computation*/ 
    if (idx < num) 
     res[idx] = src_a[idx] + src_b[idx]; 
} 

1)說例如,所執行的操作比總結要複雜得多 - 這是保證其功能的事情。我們稱之爲ComplexOp(in1,in2,out)。我將如何去實現這個函數,使vector_add_gpu()可以調用和使用它?你能舉出例子代碼嗎?

2)現在讓我們以極端的例子爲例,現在我想調用一個通用函數來處理這兩個數字。我如何設置它以便內核可以傳遞一個指向這個函數的指針並根據需要調用它?

+1

只是一個評論。這是OpenCL而不是CUDA。您不必強制使用多個工作組大小。我經常看到thouse醜陋的'如果(idx DarkZeros

回答

18

是的,這是可能的。你只需要記住OpenCL是基於C99的一些注意事項。您可以在同一個內核文件內或單獨的文件中創建其他函數,只需在開頭添加它。輔助函數不需要聲明爲內聯,但請記住,OpenCL將在調用時內聯函數。調用輔助功能時指針也不可用。

float4 hit(float4 ray_p0, float4 ray_p1, float4 tri_v1, float4 tri_v2, float4 tri_v3) 
{ 
//logic to detect if the ray intersects a triangle 
} 

__kernel void detection(__global float4* trilist, float4 ray_p0, float4 ray_p1) 
{ 
int gid = get_global_id(0); 
float4 hitlocation = hit(ray_p0, ray_p1, trilist[3*gid], trilist[3*gid+1], trilist[3*gid+2]); 
} 
+0

你是什麼意思?指示輔助功能時指針也無法使用? – Nigel

+0

在opencl內核中,指針不能在內核中的任何地方或從內核調用函數時使用。指針是這些警告之一。你不需要指針傳遞的一個原因是函數總是內聯的,而不是交給一個單獨的內存地址。 –

相關問題