求问CUDA中是否支持float4的Read\Write()???

  • 4 replies
  • 733 views
求问CUDA中是否支持float4的Read\Write()???
« 于: 十二月 03, 2021, 05:37:37 pm »
本人常年写OpenCL,最开始是写CUDA入门的,现在回归CUDA,但没有找到相关的资料说明,cuda runtime kernel当中是否可以直接进行float4的读写和计算。举个例子,OpenCL的float4 add kernel:

__kernel void Add(__global float4* a, __global float4* b, __global float4*c){
   int tid = get_global_id(0);
   c[tid] = a[tid] + b[tid]
}//OpenCL中是合法的,或者说将入参指针变为__global float* 并在相应读写时使用vload4()\vstore4()的等价操作。

但是CUDA有类似的支持吗?比如:
__global__ void Add(float4* a, float4* b, float4*c){
   int i = threadIdx.x;
   c = a + b;
}//貌似我试了下,这样子是不合法的,那请问有没有什么和OpenCL等价的方法呢???   谢谢各位!

Re: 求问CUDA中是否支持float4的Read\Write()???
« 回复 #1 于: 十二月 11, 2021, 08:11:33 pm »
程序代码: [选择]
本人常年写OpenCL,最开始是写CUDA入门的,现在回归CUDA,但没有找到相关的资料说明,cuda runtime kernel当中是否可以直接进行float4的读写和计算。举个例子,OpenCL的float4 add kernel:

__kernel void Add(__global float4* a, __global float4* b, __global float4*c){
   int tid = get_global_id(0);
   c[tid] = a[tid] + b[tid]
}//OpenCL中是合法的,或者说将入参指针变为__global float* 并在相应读写时使用vload4()\vstore4()的等价操作。

但是CUDA有类似的支持吗?比如:
__global__ void Add(float4* a, float4* b, float4*c){
   int i = threadIdx.x;
   c[i] = a[i] + b[i];
}//貌似我试了下,这样子是不合法的,那请问有没有什么和OpenCL等价的方法呢???   谢谢各位!


float4类型的读写,和能否对float4类型进行某些运算(例如你想要的加法),是两回事。

光就读写来说,float4无问题的(注意有16B的对齐要求)。

Re: 求问CUDA中是否支持float4的Read\Write()???
« 回复 #2 于: 十二月 13, 2021, 11:53:31 am »
float4类型的读写,和能否对float4类型进行某些运算(例如你想要的加法),是两回事。

光就读写来说,float4无问题的(注意有16B的对齐要求)。

我后面做了实验,确实是可以直接按照float4指针传入,另外float4 + float4的加法本身是不支持的,但支持operator overloading,所以也可以实现。  感谢[名词6]的回复!

Re: 求问CUDA中是否支持float4的Read\Write()???
« 回复 #3 于: 十二月 14, 2021, 01:29:06 pm »
我后面做了实验,确实是可以直接按照float4指针传入,另外float4 + float4的加法本身是不支持的,但支持operator overloading,所以也可以实现。  感谢[名词6]的回复!

考虑到你是OpenCL用户,注意float3的这种的区别:
sizeof(它), 12B/16B,在CUDA和OpenCL下的大小差异;
对齐性要求,4B/16B,在CUDA下和OpenCL下的对齐性要求。
其他的地方是一样的。

这是从OpenCL到CUDA用户的相互转换的时候,容易上去跪的地方。


Re: 求问CUDA中是否支持float4的Read\Write()???
« 回复 #4 于: 十二月 14, 2021, 03:06:58 pm »
考虑到你是OpenCL用户,注意float3的这种的区别:
sizeof(它), 12B/16B,在CUDA和OpenCL下的大小差异;
对齐性要求,4B/16B,在CUDA下和OpenCL下的对齐性要求。
其他的地方是一样的。

这是从OpenCL到CUDA用户的相互转换的时候,容易上去跪的地方。

明白了,感谢[名词6]指导!