#include <stdlib.h>
#include <stdio.h>
#include <cutil_inline.h>
__global__ void excute(int *data)
{ int i;
int tid=threadIdx.x;
if(tid<64)
{for(i=0;i<5;i++)
{data[tid]=data[tid]+1;
//__syncthreads();
}
}
}
int main()
{ int i,size=64;
int *h_data=(int *)malloc(sizeof(int)*size);
for(i=0;i<size;i++)
h_data[i]=2;
int *h_dataout=(int *)malloc(sizeof(int)*size);
int * d_data;
cutilSafeCall(cudaMalloc((void**)&d_data, sizeof(int)*size));
cutilSafeCall(cudaMemcpy(d_data, h_data,sizeof(int)*size,
cudaMemcpyHostToDevice));
excute<<<1,size>>>(d_data);
cutilSafeCall(cudaMemcpy(h_dataout,d_data, sizeof(int)* size, cudaMemcpyDeviceToHost));
for(i=0;i<size;i++)
printf("%d\n",h_dataout[i]);
return 0;
}问题1:该kernel是想把数组每个值循环加5次1,我让for循环中每个线程完成一次加1了再进行下一轮加1,中间设置syncthreads,结果设与不设效果一样,为什么会有这种现象?问题2:能具体说说syncthreads这个同步函数使用方法和范围吗?问题3:要从主机端向设备端传一个数怎么弄啊?比如我这里要传这个数组大小(64),是不是也要弄个指针才行啊?问题4:怎么让一个warp工作同一件事?CUDA单精度取值范围是多少啊?请各位大牛指点一下啊,小弟不胜感激
先说一下__syncthreads作用,这个同步函数的作用是让一个block内所有的线程运行到这个地方以后,才继续向下执行。为啥加上__syncthreads和不加在你的代码里反应不出来任何区别?如果你加上同步函数,那么每个线程在每次进入循环并加值以后,都等待block中所有线程都完成加值,然后再进入下一次循环,也就是说所有的warp执行的循环次数是相同的;如果不加同步函数,那么有可能有些warp执行的快,有些执行的慢(比如warp0已经全部完成5次循环,而warp99可能才完成了2次),但是当你的内核执行完,一定是所有的warp都完成了5次循环,所以运算的结果是相同的。
如果从主机向设备传一个数,和你传数组是一样的,只不过cudaMalloc的大小是1,cudaMemcpy的数据大小也是1
warp本来就是执行同一条指令的,你的意思是让warp中没有if-else的分支吧?
cuda中一个float是4字节,和CPU系统中是一样的!
1、你这里根本用不着同步,因为block内线程之间没有使用相互关联的数据。
2、同步最好不要在if条件内使用,因为同步的是block内所有线程,放在if内,很有可能某些线程不满足if,这样的话就会出现一直等待某个线程,会出现死锁。
3、一个数据可以使用指针传,但是好像没必要,直接使用函数的参数就可以了。数组的话要用指针喽
4、warp本身就是工作同一件事。取值范围取决于所使用的位数,float是4字节。