我有以下OpenCL内核，它将值从一个缓冲区复制到另一个缓冲区时，可以选择反转值("inverted"参数可以是1或-1(：-
__kernel void extraction(__global const short* src_buff, __global short* dest_buff, const int record_len, const int invert)
{
int i = get_global_id(0); // Index of record in buffer
int j = get_global_id(1); // Index of value in record
dest_buff[(i* record_len) + j] = src_buff[(i * record_len) + j] * invert;
}
源缓冲区包含一个或多个"记录"，每个记录都包含N(record_len(短值。缓冲区中的所有记录长度相等，record_len始终是32的倍数。
全局大小是2D(缓冲区中的记录数、记录长度(，我之所以选择它，是因为它似乎能最好地利用GPU并行处理，每个线程只负责复制缓冲区中一条记录中的一个值。
(顺便说一下，本地工作大小设置为NULL，允许OpenCL自己确定值(。
最近读到关于矢量的文章后，我想知道是否可以使用这些来提高性能？我理解向量的概念，但我不知道如何在实践中使用它们，部分原因是缺乏好的例子。
我确信内核的性能已经相当合理了，所以这主要是出于好奇，想看看使用向量(或其他更合适的方法(会有什么不同。
冒着有点天真的风险，我可以简单地将两个缓冲区arg类型更改为short16，并将2-D全局大小中的第二个值从"记录长度"更改为"记录长度/16"吗？这会导致每个内核线程在缓冲区之间复制一个由16个短值组成的块吗？
你的天真假设基本上是正确的，尽管你可能想向编译器添加一个提示，表明这个内核是针对向量类型优化的(规范的6.7.2节(，在你的情况下，你会添加
属性((vec_type_hint(short16((
高于您的内核函数。所以在你的例子中，你会有
__attribute__((vec_type_hint(short16)))
__kernel void extraction(__global const short16* src_buff, __global short16* dest_buff, const int record_len, const int invert)
{
int i = get_global_id(0); // Index of record in buffer
int j = get_global_id(1); // Index of value in record
dest_buff[(i* record_len) + j] = src_buff[(i * record_len) + j] * invert;
}
您的第二个全局维度应该除以16，您的record_len也应该除以16。此外，如果要指定本地大小而不是NULL，则还需要将其除以16。
不过，还有其他一些事情需要考虑
您可能认为选择最大的矢量大小应该提供最佳性能，尤其是对于这样一个简单的内核。但根据我的经验，这很少是最理想的尺寸。您可以尝试向clGetDeviceInfo询问CL_DEVICE_PREFERRED_VECTOR_WIDTH_SHORT，但对我来说，这很少是准确的(此外，它可能会给您1，这意味着编译器将尝试自动向量化，或者设备没有矢量硬件(。最好尝试不同的矢量大小，看看哪个最快。
如果您的设备支持自动向量化，并且您想尝试一下，那么删除record_len参数并将其替换为get_global_size(1(可能会有所帮助，这样编译器/驱动程序就可以将record_len除以它选择的任何向量大小。无论如何，我建议您这样做，假设record_len等于您给定该维度的全局大小。
此外，您为本地大小参数指定了NULL，以便实现自动选择大小。它保证会选择一个有效的尺寸，但不一定会选择最理想的尺寸。
最后，对于一般的OpenCL优化，您可能需要查看适用于NVIDIA硬件的NVIDIA OpenCL最佳实践指南，或适用于AMD GPU硬件的AMD APP SDK OpenCL用户指南。英伟达是2009年的，我不确定他们的硬件有多大变化。但请注意，它实际上写着：
CUDA体系结构是一种标量体系结构。因此，没有性能受益于使用矢量类型和指令。这些只能用于方便
较旧的AMD硬件(GCN之前(受益于使用矢量类型，但AMD建议不要在GCN设备上使用矢量类型(请参阅mogu的评论(。此外，如果您的目标是CPU，它将使用AVX硬件(如果可用(。