我正在尝试将c ++代码转换为Cuda代码,并且获得了以下三重嵌套的for循环,该循环将填充数组以进行进一步的OpenGL渲染(我只是在创建坐标顶点数组):
for(int z=0;z<263;z++) {
for(int y=0;y<170;y++) {
for(int x=0;x<170;x++) {
g_vertex_buffer_data_3[i]=(float)x+0.5f;
g_vertex_buffer_data_3[i+1]=(float)y+0.5f;
g_vertex_buffer_data_3[i+2]=-(float)z+0.5f;
i+=3;
}
}
}
我想获得更快的操作,因此我将使用Cuda进行上面列出的操作。我想为每个点创建一个块,并且由于每个点都有3个坐标,所以我想每个块都有3个线程。我想使用此配置,因为我有一个7600700点的3d矩阵,所以我认为最合乎逻辑的事情是创建一个由块组成的3d矩阵,然后在每个块中为x,y,z使用3个线程每个点的坐标。我将c ++代码转换为此(这是我用来理解如何使用Cuda的一个小程序,这里我只使用了几点):
__global__ void mykernel(int k, float *buffer, int size) {
const unsigned long int blockId = blockIdx.x + blockIdx.y * gridDim.x + gridDim.x * gridDim.y * blockIdx.z;
const unsigned long int threadId = (blockId * blockDim.x + threadIdx.x)*blockDim.x;
if(threadId<size ) {
buffer[threadId]=blockIdx.x+0.5;
buffer[threadId+1]=blockIdx.y+0.5;
buffer[threadId+2]=blockIdx.z+0.5;
}
}
int main(void) {
int dim=3*5*5*7;
float* g_vertex_buffer_data_2 = new float[dim];
float* g_vertex_buffer_data_3;
int i=0;
HANDLE_ERROR(cudaMalloc((void**)&g_vertex_buffer_data_3, sizeof(float)*dim));
dim3 dimBlock(3);
dim3 dimGrid(5,5,7);
mykernel<<<dimGrid, dimBlock>>>(i, g_vertex_buffer_data_3, dim);
HANDLE_ERROR(cudaMemcpy(g_vertex_buffer_data_2,g_vertex_buffer_data_3,sizeof(float)*dim,cudaMemcpyDeviceToHost));
cudaFree(g_vertex_buffer_data_3);
return 0;
}
有了这段代码,我得到的东西还不错。问题是,在if语句之后,我希望代码“跳到”下一个块,因为我得到了3次相同的结果(我有3个线程,所以代码在跳转到下一个区块)。我尝试用输出的一小部分来解释一下自己:
g_buffer_data_2[0]=0.5 g_buffer_data_2[0]=0.5
g_buffer_data_2[1]=0.5 g_buffer_data_2[1]=0.5
g_buffer_data_2[2]=0.5 g_buffer_data_2[2]=0.5
g_buffer_data_2[3]=0.5 g_buffer_data_2[3]=1.5
g_buffer_data_2[4]=0.5 g_buffer_data_2[4]=0.5
g_buffer_data_2[5]=0.5 g_buffer_data_2[5]=0.5
g_buffer_data_2[6]=0.5 g_buffer_data_2[6]=2.5
g_buffer_data_2[7]=0.5 g_buffer_data_2[7]=0.5
g_buffer_data_2[8]=0.5 g_buffer_data_2[8]=0.5
g_buffer_data_2[9]=1.5 g_buffer_data_2[9]=3.5
g_buffer_data_2[10]=0.5 g_buffer_data_2[10]=0.5
g_buffer_data_2[11]=0.5 g_buffer_data_2[11]=0.5
g_buffer_data_2[12]=1.5 g_buffer_data_2[12]=4.5
[...]
左边是我想要的东西,而右边是我想要的东西。我应该修改什么?每个块应该只使用一个线程吗?但这会降低性能吗?
由于您对CUDA不太熟悉,因此可以从构造至少dim/3
一个线程开始,其中每个线程仅填充一个点。
dim3 size(170, 170, 263);
每个块3个线程仍然太小,无法获得最佳性能。通常的选择是使用2的幂以接近设备每个块的最大线程数。warpSize
在.x
暗处使用线程是一个好习惯。线程应使用3-D块和网格进行组织,以匹配您的循环x
,y
以及z
:
dim3 dimBlock(32, 4, 4);
dim3 dimGrid((size.x + dimBlock.x - 1) / dimBlock.x,
(size.z + dimBlock.y - 1) / dimBlock.y,
(size.z + dimBlock.z - 1) / dimBlock.z);
另一方面,您的任务是float3
用来简化索引的一个好情况:
float3* g_vertex_buffer_data_3;
cudaMalloc((void**) &g_vertex_buffer_data_3,
sizeof(float3) * size.x * size.y * size.z);
所以内核应该是这样的
__global__ void mykernel(float3 *buffer, dim3 size) {
int x = blockDim.x * blockIdx.x + threadIdx.x;
int y = blockDim.y * blockIdx.y + threadIdx.y;
int z = blockDim.z * blockIdx.z + threadIdx.z;
if (x < size.x && y < size.y && z < size.z) {
float3 buf;
buf.x = x + 0.5f;
buf.y = y + 0.5f;
buf.z = -z + 0.5f;
buffer[(z * size.y + y) * size.x + x] = buf;
}
}
这就是您启动它的方式。
mykernel<<<dimGrid, dimBlock>>>(g_vertex_buffer_data_3, size);
本文收集自互联网,转载请注明来源。
如有侵权,请联系 [email protected] 删除。
我来说两句