2011-05-18 68 views
1

我为两个矩阵的元素乘法构建了一个内核,但至少在我的配置中,当每个矩阵大于2GB时,我的OpenCL内核只会更快。所以我想知道,如果是因为我的天真内核(见下文),或者是因为元素操作的本质,这意味着元素操作不会从使用GPU中获益。OpenCL中的元素操作(Cuda)

感谢您的输入!

内核:

KERNEL_CODE = """ 
// elementwise multiplication: C = A .* B. 
__kernel void matrixMul(
     __global float* C, 
     __global float* A, 
     __global float* B, 
     int width, int height) 
{ 
    // ID 
    int x = get_global_id(0); 
    int y = get_global_id(1); 

    // Multiplying 
    C[y * height + x ] = A[y * height + x] * B[y * height + x]; 
} 
""" 

附:我看过一些专家认为,CUDA与OpenCL的差异太大,无法在同一个问题中回答,因此可以免费将其从标题和标签中删除。

+0

检查在启动内核的过程中有多少时间正在丢失。你把这个比作什么?在CPU上的相同操作? – 2011-05-18 17:02:34

回答

9

这类操作有N个FLOP,但有3N个内存事务,因此它将完全受内存带宽限制。没有数据重用的余地,所以加速到参考CPU版本的上限是GPU与CPU带宽的比率。这个数量很少会超过10倍,并且可能会因将数据移入和移出GPU内存的代价而很快受到侵蚀。一般来说,这种操作最好与其他O(N)操作“融合”以提高性能。通常情况下,您永远不会只在一个内核中计算Hadamard产品,您可以将它作为一个内核中一系列O(N)操作的一部分来执行。所以,不,这不是加速的好选择,即使内核是最优的。

而你的内核肯定不是。你为每个FLOP做3个IOP,这是一个巨大的惩罚。你当然可以做一些事情来改善这一点,但是什么事情完全取决于将要运行什么类型的硬件。

2

说到元素操作:这取决于设备。例如NVidia GPU使用标量处理器(使用标量指令),不需要矢量化。相反,ATI采用5d(或4d)VLIW处理器,对于这些处理器来说,矢量化至关重要。但是,编译器有时可以执行它,而不是直接在代码中使用矢量数据类型,但是在为ATI的GPU进行优化时首先要做的事情。不过,正如talonmies所指出的那样,上述算法几乎没有内存带宽限制,并且不能仅仅使用GPU等待很多加速。

0

你发布的内核应该至少和CPU一样快。但是你根本没有使用合并内存访问!

这是杀死你的表现。

但是,正如@talonmies所述。对于GPU来说这不是一个好例子。你在内存拷贝中丢失了所有的时间。