Mvapich在内核运行时在CUDA内存上僵局

我尝试获得一个与mvapich cuda8一起工作的MPI-CUDA程序。我以前确实使用OpenMPI成功运行了该程序，但是我想测试是否在Mvapich获得更好的性能。不幸的是，如果使用MVAPICH时CUDA内核同时运行。

我下载了mvapich2-2.2并用配置标志

启用MPI调用CUDA内存。Mcast被禁用，因为我无法没有标志。

我在运行应用程序之前使用以下标志：

export MV2_USE_CUDA=1
export MV2_GPUDIRECT_GDRCOPY_LIB=/path/to/gdrcopy/
export MV2_USE_GPUDIRECT=1

mpi_isend/recv同时运行cuda内核时工作正常。但是在我的程序中，MPI在内核正在运行时将数据发送并接收到GPU内存的数据很重要。

我提出了这一行为的两个可能原因。首先，由于某些原因，Mvapich试图运行自己的CUDA内核来从GPU内存发送数据，并且由于GPU已经充分利用，因此该内核不会被安排。第二种可能性：Mvapich使用cudamemcpy某处（不是异步版本），该版本会阻止内核完成执行。

有人可以确认我的假设之一吗？Mvapich中是否有一个我不知道的问题？

编辑：

在这里，一个" simpel"代码说明了我的问题。使用OpenMPI执行代码时，它可以正确执行并终止。使用mvapich2，它在标记的mpi_send函数处僵硬。

#include <stdio.h>
#include <stdlib.h>
#include <cuda.h>
#include <mpi.h>

__global__ void kernel(double * buffer, int rank)
{
    volatile double *buf = buffer;
    if(rank == 0){
        while(buf[0] != 3){}
    } else {
        while(buf[0] != 2){}
    }
}

int main(int argc, char **argv)
{
    double host_buffer[1];
    MPI_Init(&argc, &argv);
    int world_size, world_rank;
    MPI_Comm_size(MPI_COMM_WORLD, &world_size);
    MPI_Comm_rank(MPI_COMM_WORLD, &world_rank);
    printf("Im rank %dn", world_rank);
    cudaSetDevice(world_rank);
    double * dev_buffer;
    cudaError_t err = cudaMalloc(&dev_buffer, sizeof(double));
    if(world_rank == 0){
        host_buffer[0] = 1;
        cudaError_t err = cudaMemcpy(dev_buffer, host_buffer, sizeof(double), cudaMemcpyHostToDevice);
        MPI_Send(dev_buffer, 1, MPI_DOUBLE, 1, 0, MPI_COMM_WORLD);
        printf("[%d]First send does not deadlockn", world_rank);
    }else {
        MPI_Recv(dev_buffer, 1, MPI_DOUBLE, 0, 0, MPI_COMM_WORLD, MPI_STATUS_IGNORE);
        printf("[%d]Received first messagen", world_rank);
    }
    cudaStream_t stream, kernel_stream;
    cudaStreamCreate(&stream);
    cudaStreamCreate(&kernel_stream);
    printf("[%d]launching kerneln", world_rank);
    kernel<<<208, 128, 0, kernel_stream>>>(dev_buffer, world_rank);
    if(world_rank == 0){
        //rank 0
        host_buffer[0] = 2;
        cudaMemcpyAsync(
            dev_buffer, host_buffer, sizeof(double),
            cudaMemcpyHostToDevice,
            stream
        );
        cudaStreamSynchronize(stream);
        printf("[%d]Send messagen", world_rank);
        MPI_Send(dev_buffer, 1, MPI_DOUBLE, 1, 0, MPI_COMM_WORLD); //mvapich2 deadlocks here
        printf("[%d]Message sentn", world_rank);
        printf("[%d]Receive messagen", world_rank);
        MPI_Recv(dev_buffer, 1, MPI_DOUBLE, 1, 0, MPI_COMM_WORLD, MPI_STATUS_IGNORE);
        printf("[%d]Message receivedn", world_rank);
        cudaStreamSynchronize(kernel_stream);
        printf("[%d]kernel finishedn", world_rank);
    } else {
        //rank 1
        printf("[%d]Receive messagen", world_rank);
        MPI_Recv(dev_buffer, 1, MPI_DOUBLE, 0, 0, MPI_COMM_WORLD, MPI_STATUS_IGNORE);
        printf("[%d]Message receivedn", world_rank);
        cudaStreamSynchronize(kernel_stream);
        printf("[%d]kernel finishedn", world_rank);
        host_buffer[0] = 3;
        cudaMemcpyAsync(
            dev_buffer, host_buffer, sizeof(double),
            cudaMemcpyHostToDevice,
            stream
        );
        cudaStreamSynchronize(stream);
        printf("[%d]Send messagen", world_rank);
        MPI_Send(dev_buffer, 1, MPI_DOUBLE, 0, 0, MPI_COMM_WORLD);
        printf("[%d]Message sentn", world_rank);
    }
    printf("[%d]Stopped executionn", world_rank);
    MPI_Finalize();
}