在没有内存移动的情况下交换 CUDA Thrust 设备向量

Question

如果我有两个 cudaMalloced 数组，我可以通过简单地交换相关指针来交换它们而无需内存移动。

如果我有两个 CUDA Thrust device_vectors，比如 d_a 和 d_b，我可以使用第三个临时向量交换它们，比如 d_c，但是这个将需要记忆运动。

我的问题是：有没有办法在不移动内存的情况下交换 CUDA Thrust device_vectors？

Answer 1

我不知道。

没有暴露的构造函数接受现有的 device_ptr，并且 device_vector 中的底层基向量是私有的，因此无法深入研究并自己执行指针交换。这些将是我能想到的在不触发标准复制构造函数的情况下完成这项工作的唯一方法。

编辑补充说这个答案似乎是错误的。似乎最近（可能在 thrust 1.6 左右）的变化已经实现了一个内部指针交换交换机制，可以通过 device_vector.swap() 调用。这绕过了 swap() 的常用复制构造函数习惯用法，并且不会触发内存传输 .

Answer 2

看来device_vector.swap()避免了记忆移动。

确实，请考虑以下代码：

#include "cuda_runtime.h"
#include "device_launch_parameters.h"

#include <stdio.h>

#include <thrust\device_vector.h>

void printDeviceVector(thrust::device_vector<int> &d_a) {

    for (int k = 0; k < d_a.size(); k++) {

        int temp = d_a[k];
        printf("%i\n", temp);

    }

}

int main()
{
    const int N = 10;

    thrust::device_vector<int> d_a(N, 1);
    thrust::device_vector<int> d_b(N, 2);

    // --- Original
    printf("Original device vector d_a\n");
    printDeviceVector(d_a);
    printf("Original device vector d_b\n");
    printDeviceVector(d_b);

    d_b.swap(d_a);

    // --- Original
    printf("Final device vector d_a\n");
    printDeviceVector(d_a);
    printf("Final device vector d_b\n");
    printDeviceVector(d_b);

    d_a.clear();
    thrust::device_vector<int>().swap(d_a); 
    d_b.clear();
    thrust::device_vector<int>().swap(d_b);

    cudaDeviceReset();

    return 0;
}

使用

    d_b.swap(d_a);

如果我们分析它，我们会在时间轴中看到没有设备到设备的内存移动：

如果在另一边，我们将d_b.swap(d_a)改为

d_b = d_a;

然后设备到设备的移动出现在时间轴中：

最后，时机明显有利于 d_b.swap(d_a)，而不是 d_b = d_a。对于N = 33554432，时间是

d_b.swap(d_a)     0.001152ms
d_b = d_a         3.181824ms

在没有内存移动的情况下交换 CUDA Thrust 设备向量

Swapping CUDA Thrust device vectors without memory movements

cuda

thrust