CUDA多GPU编程系列一

2013-11-11 By SuHui

概述

前言及声明

为什么需要多GPU?

GPU通信(Single Host, Multiple GPUs)

单个CPU线程管理多GPU

    
    cudaSetDevice( 0 ); 
    kernel<<<...>>>(...); 
    cudaMemcpyAsync(...); 
    cudaSetDevice( 1 ); 
    kernel<<<...>>>(...); 

统一地址(CUDA4.0及以后)

UVA和多GPU编程

Peer-to-peer内存拷贝

一个实例:4个GPU 从左到右传输法

GPU 那么我们如何实现这四个GPU之间的数据交换呢? 第一步:send "right"/receive from "left" GPU 第二步:send "left"/ receive from "right" GPU 这步中三次传输是并发的,效果惊人,吞吐最大达到15GB/s

for( int i=0; i <num_gpus-1; i++ ) // "right" stage 
cudaMemcpyPeerAsync( d_a[i+1], gpu[i+1], d_a[i], gpu[i], num_bytes, stream[i] ); 
for( int i=0; i <num_gpus; i++ ) 
cudaStreamSynchronize( stream[i] ); 
for( int i=1; i <num_gpus; i++ ) // "left" stage 
cudaMemcpyPeerAsync(d_b[i-1], gpu[i-1], d_b[i], gpu[i], num_bytes, stream[i] );

GPU通信(Multiple Host, Multiple GPUs)

不同节点间GPU的通信

    cudaMemcpyAsync( ..., stream_halo[i] ); 
    cudaStreamSynchronize( stream_halo[i] ); 
    MPI_Sendrecv( ... ); 
    cudaMemcpyAsync( ..., stream_halo[i] ); 

多GPU,stream,events

参考之前做的pdf

两个有用的API

cudalpc* API

    //Process A: 
    cudaIpcMemHandle_t   handle_a; 
    cudaIpcGetMemHandle( &handle_a, (void*)sender_dev_buffer ); 
    //Process B: 
    cudaIpcMemHandle_t   handle_b;
    MPI_Irecv(handle_b,sizeof(cudaIpcMemHandle_t),...)
    cudaIpcOpenMemHandle( b, handle_b,   cudaIpcMemLazyEnablePeerAccess ); 
    // 这里使用设备缓冲(device buffer)b
    cudaIpcCloseMemHandle( handle_b ); 

其实就是刚才的MPI

点击查看评论