没有分配缓冲区内存意味着您必须保留复制顺序,不能并行以充分利用 GPU 硬件。
这是使用 Thrust 和缓冲内存执行此操作的版本。
它需要 Thrust 1.6.0+,因为迭代器上使用了 lambda 表达式仿函数。
#include "thrust/device_vector.h"
#include "thrust/iterator/counting_iterator.h"
#include "thrust/iterator/permutation_iterator.h"
#include "thrust/iterator/transform_iterator.h"
#include "thrust/copy.h"
#include "thrust/functional.h"
using namespace thrust::placeholders;
int main()
{
const int oldChunk = 100, newChunk = 95;
const int size = 10000;
thrust::device_vector<float> v(
thrust::counting_iterator<float>(0),
thrust::counting_iterator<float>(0) + oldChunk * size);
thrust::device_vector<float> buf(newChunk * size);
thrust::copy(
thrust::make_permutation_iterator(
v.begin(),
thrust::make_transform_iterator(
thrust::counting_iterator<int>(0),
_1 / newChunk * oldChunk + _1 % newChunk)),
thrust::make_permutation_iterator(
v.begin(),
thrust::make_transform_iterator(
thrust::counting_iterator<int>(0),
_1 / newChunk * oldChunk + _1 % newChunk))
+ buf.size(),
buf.begin());
return 0;
}
我认为由于使用了 mod operator ,上述版本可能无法达到最高性能%
。为了获得更高的性能,您可以考虑使用 cuBLAS 函数cublas_geam()
float alpha = 1;
float beta = 0;
cublasSgeam(handle, CUBLAS_OP_N, CUBLAS_OP_N,
newChunk, size,
&alpha,
thrust::raw_pointer_cast(&v[0]), oldChunk,
&beta,
thrust::raw_pointer_cast(&v[0]), oldChunk,
thrust::raw_pointer_cast(&buf[0]), newChunk);