在GPU中查找最低速度比CPU慢

我已經實現了此代碼：http://www.cuvilib.com/Reduction.pdf以便計算矩陣元素的總和。在GPU中查找最低速度比CPU慢

但是，在GPU中，它的運行速度比在CPU中慢得多。

我得到了i7處理器和NVIDIA GT 540M顯卡。

它應該是這樣或其他什麼？

編輯：我在Ubuntu 13.04中使用上述代碼的第3版，並使用Eclipse Nsight進行編譯。矩陣的大小是2097152個元素。它在3.6 ms內執行，而CPU版本在1.0 ms左右。下面是整個代碼：

#include <stdio.h> 
#include <stdlib.h> 
#include <thrust/sort.h> 
#include <sys/time.h> 
#include <omp.h> 
#include <iostream> 
#include <algorithm> 

#define MIN(a,b) (((a)<(b))?(a):(b)) 



static const int WORK_SIZE = 2097152; 



int find_min(int *a,int length){ 
    int min = a[0]; 
    for (int i=1;i<length;i++) 
      if (a[i]<min) 
     min=a[i]; 
    return min; 
} 


__global__ static void red_min(int *g_idata,int *g_odata) { 
    extern __shared__ int sdata[]; 
    unsigned int tid = threadIdx.x; 
    unsigned int i = blockIdx.x * blockDim.x + threadIdx.x; 
    sdata[tid]= g_idata[i]; 
    __syncthreads(); 

    for(unsigned int s=blockDim.x/2; s > 0; s >>= 1) { 
     if (tid<s) { 
      sdata[tid] = MIN(sdata[tid],sdata[tid + s]); 
     } 
     __syncthreads(); 
    } 
    if (tid == 0) 
     g_odata[blockIdx.x] = sdata[0]; 
} 





int main(void) { 
    int *d1,*d2; 
    int i,*result; 
    int *idata,*fdata; 
    srand (time(NULL)); 
    result = (int *)malloc(sizeof(int)); 
    idata = (int *)malloc(WORK_SIZE*sizeof(int)); 
    fdata = (int *)malloc(WORK_SIZE*sizeof(int)); 
    cudaMalloc((int**)&d1,WORK_SIZE*sizeof(int)); 
    cudaMalloc((int**)&d2,WORK_SIZE*sizeof(int)); 


    for (i = 0; i < WORK_SIZE; i++){ 
     idata[i] = rand(); 
     fdata[i] = i; 
    } 
    struct timeval begin, end; 
    gettimeofday(&begin, NULL); 
    *result = find_min(idata,WORK_SIZE); 
    printf("Minimum Element CPU: %d \n", *result); 
    gettimeofday(&end, NULL); 
    int time = (end.tv_sec * (unsigned int)1e6 + end.tv_usec) - (begin.tv_sec * (unsigned int)1e6 + begin.tv_usec); 
    printf("Microseconds elapsed CPU: %d\n", time); 

    cudaMemcpy(d1,idata,WORK_SIZE*sizeof(int),cudaMemcpyHostToDevice); 



    cudaEvent_t start, stop; 
    cudaEventCreate(&start); 
    cudaEventCreate(&stop); 
    cudaEventRecord(start,0); 
    int num_blocks = 16384; 
    bool flag = true; 
    while (num_blocks>0){ 
     if (flag) { 
      red_min<<<num_blocks,128,128*sizeof(int)>>>(d1,d2); 
     } 
     else { 
      red_min<<<num_blocks,128,128*sizeof(int)>>>(d2,d1); 
     } 
     num_blocks /= 128; 
     flag = !flag; 
}

來源

2013-11-14 user2424276

該演示文稿中有許多不同版本的代碼。我們不知道您正在處理的是什麼尺寸的陣列。像編譯代碼的許多細節也會影響性能。您沒有提及您使用的是哪種CUDA版本，以及您是在Linux還是Windows。儘管您將該演示文稿鏈接起來，但您幾乎沒有提供任何信息。顯示完整的代碼，顯示如何編譯它，並顯示實際測量結果和輸出。然後有人可以幫助你。 –

請提供您的*完整*代碼。我希望看到一切。這不清楚嗎？ –

@ user2424276使用分析器。 – gpuguy

GT540M是移動GPU，所以我假設你在筆記本電腦上運行，而且您可能會被託管在540M GPU的X顯示。

我建你的代碼的完整版本：

#include <stdio.h> 
#include <stdlib.h> 
#include <thrust/sort.h> 
#include <sys/time.h> 
#include <omp.h> 
#include <iostream> 
#include <algorithm> 

#define MIN(a,b) (((a)<(b))?(a):(b)) 



static const int WORK_SIZE = 2097152; 



int find_min(int *a,int length){ 
    int min = a[0]; 
    for (int i=1;i<length;i++) 
      if (a[i]<min) 
     min=a[i]; 
    return min; 
} 


__global__ static void red_min(int *g_idata,int *g_odata) { 
    extern __shared__ int sdata[]; 
    unsigned int tid = threadIdx.x; 
    unsigned int i = blockIdx.x * blockDim.x + threadIdx.x; 
    sdata[tid]= g_idata[i]; 
    __syncthreads(); 

    for(unsigned int s=blockDim.x/2; s > 0; s >>= 1) { 
     if (tid<s) { 
      sdata[tid] = MIN(sdata[tid],sdata[tid + s]); 
     } 
     __syncthreads(); 
    } 
    if (tid == 0) 
     g_odata[blockIdx.x] = sdata[0]; 
} 





int main(void) { 
    int *d1,*d2; 
    int i,*result; 
    int *idata,*fdata; 
    srand (time(NULL)); 
    result = (int *)malloc(sizeof(int)); 
    idata = (int *)malloc(WORK_SIZE*sizeof(int)); 
    fdata = (int *)malloc(WORK_SIZE*sizeof(int)); 
    cudaMalloc((int**)&d1,WORK_SIZE*sizeof(int)); 
    cudaMalloc((int**)&d2,WORK_SIZE*sizeof(int)); 


    for (i = 0; i < WORK_SIZE; i++){ 
     idata[i] = rand(); 
     fdata[i] = i; 
    } 
    struct timeval begin, end; 
    gettimeofday(&begin, NULL); 
    *result = find_min(idata,WORK_SIZE); 
    printf("Minimum Element CPU: %d \n", *result); 
    gettimeofday(&end, NULL); 
    int time = (end.tv_sec * (unsigned int)1e6 + end.tv_usec) - (begin.tv_sec * (unsigned int)1e6 + begin.tv_usec); 
    printf("Microseconds elapsed CPU: %d\n", time); 

    cudaMemcpy(d1,idata,WORK_SIZE*sizeof(int),cudaMemcpyHostToDevice); 



    cudaEvent_t start, stop; 
    cudaEventCreate(&start); 
    cudaEventCreate(&stop); 
    cudaEventRecord(start,0); 
    int num_blocks = 16384; 
    bool flag = true; 
    int loops = 0; 
    while (num_blocks>0){ 
     if (flag) { 
      red_min<<<num_blocks,128,128*sizeof(int)>>>(d1,d2); 
     } 
     else { 
      red_min<<<num_blocks,128,128*sizeof(int)>>>(d2,d1); 
     } 
     num_blocks /= 128; 
     flag = !flag; 
     loops++; 
    } 
    cudaEventRecord(stop, 0); 
    cudaEventSynchronize(stop); 
    float et = 0.0f; 
    cudaEventElapsedTime(&et, start, stop); 
    printf("GPU time: %fms, in %d loops\n", et, loops); 
    int gpuresult; 
    if (flag) 
     cudaMemcpy(&gpuresult, d1, sizeof(int), cudaMemcpyDeviceToHost); 
    else 
     cudaMemcpy(&gpuresult, d2, sizeof(int), cudaMemcpyDeviceToHost); 
    printf("GPU min: %d\n", gpuresult); 
    return 0; 
}

編譯它：

$ nvcc -O3 -arch=sm_20 -o t264 t264.cu

並運行它在M2050 GPU，RHEL 5.5，CUDA 5.5，至強X5650 CPU

$ ./t264 
Minimum Element CPU: 288 
Microseconds elapsed CPU: 1217 
GPU time: 0.621408ms, in 3 loops 
GPU min: 288 
$

所以我的CPU結果與您的結果非常接近，但我的GPU結果大約快了5-6倍。如果我們將M2050與GT540M進行比較，我們可以看到M2050有14個SM，而GT540M有2個。更重要的是，M2050的內存帶寬約爲GT540M GPU的5倍（GT540M的理論峯值爲28.8GB/s，而大約150GB/s峯值理論爲M2050）

由於寫得很好的並行減少是GPU上的內存帶寬受限代碼，所以GPU和我的GPU之間的速度差異是有道理的。

所以我會說你的結果可能是關於預期的，爲了獲得更好的結果，你可能需要更快的GPU。另外，如果您的GT540M還承載了X顯示器，則顯卡活動可能會損壞GPU時序。如果我們計算單個內核，這通常不是問題 - 內核執行會暫時中斷顯示處理。但是當我們連續對內核序列進行計時時，顯示任務可能會跳入並在內核調用之間執行（當GPU被要求同時支持顯示器並處理CUDA代碼時，GPU是多任務的）。因此，這可能會對您的情況產生影響。

來源

2013-11-14 23:12:11

謝謝你的回答！ – user2424276

很好的比較，以瞭解性能 – pQB

在GPU中查找最低速度比CPU慢

回答

相關問題