CudaMemcpy 오버 헤드를 줄이는 방법

Question

배열의 크기가 3000이고 배열에 0과 1.i가 포함되어 있습니다.이 배열을 0 번째 index.i부터 시작하여 해당 위치에 저장된 첫 번째 배열 위치를 찾고 싶습니다.이 배열을 Host 및이 ​​배열로 전송하십시오. 내가 host.in 내 프로그램에서 색인을 순차적으로 계산 한 device.so에서 계산됩니다.이 계산을 반복적으로 4000 번 이상하고 싶습니다 .i이 process.is에 걸리는 시간을 줄이려면 다른 방법으로이 작업을 수행 할 수 있습니다. 이 배열은 실제로 GPU에서 계산되므로 매번 전송해야합니다.

int main() 
{ 
for(int i=0;i<4000;i++) 
{ 
    cudaMemcpy(A,dev_A,sizeof(int)*3000,cudaMemcpyDeviceToHost); 
    int k; 
    for(k=0;k<3000;k++) 
    { 
     if(A[k]==1) 
     { 
      break; 
     } 
    } 
    printf("got k is %d",k); 
} 
} 

전체 코드는이 사용법 #include "cuda.h" #DEFINE 크기 2688 #DEFINE 블록 (14) #DEFINE 192

__global__ void kernel(int *A,int *d_pos) 
{ 
int thread_id=threadIdx.x+blockIdx.x*blockDim.x; 
while(thread_id<SIZE) 
{ 
    if(A[thread_id]==INT_MIN) 
    { 
     *d_pos=thread_id; 
     return; 
    } 
    thread_id+=1; 
} 

}

__global__ void kernel1(int *A,int *d_pos) 
{ 
int thread_id=threadIdx.x+blockIdx.x*blockDim.x; 
if(A[thread_id]==INT_MIN) 
{ 
    atomicMin(d_pos,thread_id); 
} 

스레드 #INCLUDE 같다

}

int main() 
{ 
int pos=INT_MAX,i; 
int *d_pos; 
int A[SIZE]; 
int *d_A; 
for(i=0;i<SIZE;i++) 
{ 
    A[i]=78; 
} 
A[SIZE-1]=INT_MIN; 
cudaMalloc((void**)&d_pos,sizeof(int)); 
cudaMemcpy(d_pos,&pos,sizeof(int),cudaMemcpyHostToDevice); 
cudaMalloc((void**)&d_A,sizeof(int)*SIZE); 
cudaMemcpy(d_A,A,sizeof(int)*SIZE,cudaMemcpyHostToDevice); 

cudaEvent_t start_cp1,stop_cp1; 
    cudaEventCreate(&stop_cp1); 
    cudaEventCreate(&start_cp1); 
    cudaEventRecord(start_cp1,0); 

kernel1<<<BLOCKS,THREADS>>>(d_A,d_pos); 

cudaEventRecord(stop_cp1,0); 
    cudaEventSynchronize(stop_cp1); 
    float elapsedTime_cp1; 
    cudaEventElapsedTime(&elapsedTime_cp1,start_cp1,stop_cp1); 
    cudaEventDestroy(start_cp1); 
    cudaEventDestroy(stop_cp1); 
    printf("\nTime taken by kernel is %f\n",elapsedTime_cp1); 
cudaDeviceSynchronize(); 

cudaEvent_t start_cp,stop_cp; 
    cudaEventCreate(&stop_cp); 
    cudaEventCreate(&start_cp); 
    cudaEventRecord(start_cp,0); 

cudaMemcpy(A,d_A,sizeof(int)*SIZE,cudaMemcpyDeviceToHost); 

    cudaEventRecord(stop_cp,0); 
    cudaEventSynchronize(stop_cp); 
    float elapsedTime_cp; 
    cudaEventElapsedTime(&elapsedTime_cp,start_cp,stop_cp); 
    cudaEventDestroy(start_cp); 
    cudaEventDestroy(stop_cp); 
    printf("\ntime taken by copy of an array is %f\n",elapsedTime_cp); 






    cudaEvent_t start_cp2,stop_cp2; 
    cudaEventCreate(&stop_cp2); 
    cudaEventCreate(&start_cp2); 
    cudaEventRecord(start_cp2,0); 

    cudaMemcpy(&pos,d_pos,sizeof(int),cudaMemcpyDeviceToHost); 

    cudaEventRecord(stop_cp2,0); 
    cudaEventSynchronize(stop_cp2); 
    float elapsedTime_cp2; 
    cudaEventElapsedTime(&elapsedTime_cp2,start_cp2,stop_cp2); 
    cudaEventDestroy(start_cp2); 
    cudaEventDestroy(stop_cp2); 
    printf("\ntime taken by copy of a variable is %f\n",elapsedTime_cp2); 


cudaMemcpy(&pos,d_pos,sizeof(int),cudaMemcpyDeviceToHost); 
printf("\nminimum index is %d\n",pos); 
return 0; 
} 

성능 향상을 위해이 코드의 총 소요 시간을 줄이려면 어떻게해야합니까?

Answer 1

GPU에서 4000 번 커널을 실행하는 경우 다른 스트림을 통해 커널에서 비동기 실행을 사용해야 할 수 있습니다. cudaMemCpyAsync를 사용하면 호스트의 논 블로킹 (non-blocking) 기능을 사용하는 것이 더 빠를 수도 있습니다 (커널의 M 배를 실행하는 경우).

빠른 도입 스트리밍 및 비동기 실행하기 : https://devblogs.nvidia.com/parallelforall/how-overlap-data-transfers-cuda-cc/

스트림과 동시성이 도울 수 http://on-demand.gputechconf.com/gtc-express/2011/presentations/StreamsAndConcurrencyWebinar.pdf

희망 ... 내용을 생산하는 커널의 속도가