第一次从推力执行排序需要太长时间

如何解决第一次从推力执行排序需要太长时间

我正在使用 opengl（实现 sph 算法）开发流体模拟器。我尝试了很多方法来运行我的模拟器，首先我使用八叉树，然后是哈希图，现在我尝试使用 Z 顺序，为此我需要根据它们的索引对我的粒子进行排序。

让我有点理解的是，如果我有一个推力::排序需要 15 毫秒，如果我有两个推力::排序需要 17 毫秒。

为了更清楚地说明，我正在 opengl 中做我的模拟器（我所有的缓冲区都是使用 opengl 创建的），我使用 cuda interops 以便使用 cuda 对我的缓冲区进行排序。

这是我获取缓冲区并将它们“链接”到 cuda 的部分

//I use this if to do the registerBuffer only one time
if (first == 0) {
        //index
        IBuffer* bIndex = RESOURCEMANAGER->getBuffer("particleLib::Index");
        int buffIdindex = bIndex->getPropi(IBuffer::ID);
        //Position
        IBuffer* bPosition = RESOURCEMANAGER->getBuffer("particleLib::Position");
        int buffIdPosition = bPosition->getPropi(IBuffer::ID);
        //TempIndex
        IBuffer* bTempIndex = RESOURCEMANAGER->getBuffer("particleLib::TempIndex");
        int buffIdTempIndex = bTempIndex->getPropi(IBuffer::ID);
        //VeLocity
        IBuffer* bVeLocity = RESOURCEMANAGER->getBuffer("particleLib::VeLocity");
        int buffIdVeLocity = bVeLocity->getPropi(IBuffer::ID);

        // register this buffer object with CUDA
        //So devia chamar isto uma vez
        cudaGraphicsglregisterBuffer(&cuda_ssbo_Index,buffIdindex,cudaGraphicsMapFlagsNone);
        cudaGraphicsglregisterBuffer(&cuda_ssbo_TempIndex,buffIdTempIndex,cudaGraphicsMapFlagsNone);
        cudaGraphicsglregisterBuffer(&cuda_ssbo_Position,buffIdPosition,cudaGraphicsMapFlagsNone);
        cudaGraphicsglregisterBuffer(&cuda_ssbo_VeLocity,buffIdVeLocity,cudaGraphicsMapFlagsNone);
        first = 1;
    }
    

    // map OpenGL buffer object for writing from CUDA
    int* dptRSSboIndex;
    int* dptRSSboTempIndex;
    float4 * dptRSSboPosition;
    float4 * dptRSSboVeLocity;

    cudaGraphicsMapResources(1,&cuda_ssbo_Index,0);
    cudaGraphicsMapResources(1,&cuda_ssbo_TempIndex,&cuda_ssbo_Position,&cuda_ssbo_VeLocity,0);

    size_t num_bytesssbo_Index;
    size_t num_bytesssbo_TempIndex;
    size_t num_bytesssbo_Position;
    size_t num_bytesssbo_VeLocity;

    cudaGraphicsResourceGetMappedPointer((void**)&dptRSSboIndex,&num_bytesssbo_Index,cuda_ssbo_Index);
    cudaGraphicsResourceGetMappedPointer((void**)&dptRSSboTempIndex,&num_bytesssbo_TempIndex,cuda_ssbo_TempIndex);
    cudaGraphicsResourceGetMappedPointer((void**)&dptRSSboPosition,&num_bytesssbo_Position,cuda_ssbo_Position);
    cudaGraphicsResourceGetMappedPointer((void**)&dptRSSboVeLocity,&num_bytesssbo_VeLocity,cuda_ssbo_VeLocity);

    mysort(&dptRSSboIndex,&dptRSSboPosition,&dptRSSboTempIndex,&dptRSSboVeLocity,216000);

    cudaGraphicsUnmapResources(1,0);
    cudaGraphicsUnmapResources(1,0);

这是mysort的代码

void mysort(int ** index1,float4 ** values1,int** index2,float4 ** values2,int particles){
    
    thrust::device_ptr<int> i1buff = thrust::device_pointer_cast(*(index1));
    thrust::device_ptr<float4> v1buff = thrust::device_pointer_cast(*(values1));
    thrust::device_ptr<int> i2buff = thrust::device_pointer_cast(*(index2));
    thrust::device_ptr<float4> v2buff = thrust::device_pointer_cast(*(values2));

    //sorts
    thrust::stable_sort_by_key(i1buff,i1buff + particles,v1buff); // 15 ms
    //cudaThreadSynchronize();
    thrust::stable_sort_by_key(i2buff,i2buff + particles,v2buff); // 17 ms


    //repetido so para ver o tempo
    thrust::stable_sort_by_key(i1buff,v1buff);
    //cudaThreadSynchronize();
    thrust::stable_sort_by_key(i2buff,v2buff); //4 sorts -> 19 ms

    //cudaThreadSynchronize();
}

有人能解释一下这是怎么回事吗？

编辑1：我使用 cudaDeviceSynchronize() 来测量每次排序所需的时间（如@Jérôme-Richard 所示），并且第一次排序总是需要更长的时间，即使我更改订单。另一个事实是，如果我的相机靠近场景，第一种需要更长的时间，这表明 Cuda 可能正在等待 opengl 完成他的工作，使第一种“需要更长的时间”。我也尝试在 mysort() 函数上不排序，我里面唯一的东西是 cudaDeviceSynchronize() 并且花了 15 毫秒，这再次表明 cuda 可能在 opengl 上完成最后一帧的工作.

编辑2：我做了更多的调试，我认为似乎是真的。真正的减速来自 cudaGraphicsMapResources 调用。据此（cudaGraphicsMapResources slow speed when mapping DirectX texture）：

此函数提供同步保证，在 cudaGraphicsMapResources() 之前发出的任何图形调用将在流中发出的任何后续 CUDA 工作开始之前完成。

所以是的，它正在等待opengl绘制一些东西，因为camara距离影响cudaGraphicsMapResources花费的时间。

解决方法

有两点可以解释您的观察：

第一个 CUDA 函数调用隐式初始化运行时（相当慢）。
要排序的数组的实际内容可能/通常会影响排序的性能（关于 Thrust 实现中使用的算法）。数据排序后，可以更快地排序，因为它们已经排序。
Thrust 在许多提供的函数中进行很少的同步（即它调用 cudaDeviceSynchronize），以确保可以从 CPU 端安全地读取从 GPU 传输的返回数据。当关于计算数据的结果提交多个相互依赖的 CUDA 内核时，它还在内部使用这种同步（您可以使用 Nvidia 分析器看到这一点）。对于在此函数之前进行的先前异步 CUDA 调用，过度同步会增加不必要的开销。