使用cudaEvent的時機****（）VS clock_gettime（）

Question

我試圖計時我的代碼。我被告知可以應用cudaEvent ****（）。在此期間，我的原始代碼使用clock_gettime（）來計時。我按如下方式打印由cudaEvent ****（）和clock_gettime（）測量的結果。這是我真的很困惑。

通過cudaEvent測量****（）

• INIT數據結構：1971.517578ms
• 建立上下文：0.007296ms
• 重新排列數據：234.271423ms
• 拷貝數據：53.402176ms
• 時間步進：17221.333984ms

measur由clock_gettime（）

• 初始化數據結構ED：1.802874s
• 建立上下文：20.541891s
• 重新排列數據：0.235464s
• 拷貝數據：0.051851s
• 時間步：8.429955s

注：

• init數據結構：完全在CPU上工作
• 建立上下文：僅一行：cudaFree（（void *）0）;
• 重新排列數據：在CPU完全工作
• 複製數據：從主機傳送數據到設備
• 時間步：兩個內核函數涉及

Q1：測量用過的「建立上下文」的時間通過cudaEvent ****（）（0.0072ms）與clock_gettime（）（〜20.5s）測量的完全不同。實際上，這部分只有一條線建立了一個上下文。 cudaFree（0） 這種巨大的差異是如何發生的？ Q2：由cudaEvent ****（）（〜17.221s）測量的「時間步進」花費的時間是clock_gettime（）（〜8.43s）測量的時間的兩倍。有人告訴我，異步可能是一個可能的原因，但我並不真正瞭解它。任何人都可以幫助我解決它嗎？

Q3：花費的掛鐘時間非常接近clock_gettime（）測量的時間。但是，我被告知cudaEvent ****（）在cuda代碼的定時中更爲可取。我不知道我應該選擇哪一個。

===============================更新============== ===================== 以下是我的代碼中的一部分，其中定義了一些定時函數和宏。

#define TIMING 1 
#if TIMING 
double get_time() 
{ 
    struct timespec time; 
    clock_gettime(CLOCK_REALTIME, &time); 
    return (double)time.tv_sec + (double)time.tv_nsec * 1.0e-9 ; 
} 
#endif 
#define CUDATIMING 0 
#if CUDATIMING 
#define cuda_timing_init \ 
    cudaEvent_t startEvent, stopEvent;\ 
    float timeEvent;\ 
    cudaEventCreate(&startEvent);\ 
    cudaEventCreate(&stopEvent); 
#define cuda_timing_begin \ 
    cudaEventRecord(startEvent, 0); 
#define cuda_timing_stop(str) \ 
    cudaEventRecord(stopEvent, 0);\ 
    cudaEventSynchronize(stopEvent);\ 
    cudaEventElapsedTime(&timeEvent, startEvent, stopEvent);\ 
    printf("time spent of %s: %fms\n", str, timeEvent); 
#define cuda_timing_destroy \ 
    cudaEventDestroy(startEvent);\ 
    cudaEventDestroy(stopEvent); 
#endif 

我使用這些函數和宏來計時。

===========================更新20150823 ================= ==============

這裏是我的代碼的基本結構，包括時間。我不確定它是否有助於解決我的時間問題。

void 
copy_float_from_host_to_dev(float *h_p, float **d_pp, int size) 
{ 
    if_error(cudaMalloc(d_pp, size)); 
    if_error(cudaMemcpy(*d_pp, h_p, size, cudaMemcpyHostToDevice)); 
} 

void 
copy_int_from_host_to_dev(int *h_p, int **d_pp, int size) 
{ 
    if_error(cudaMalloc(d_pp, size)); 
    if_error(cudaMemcpy(*d_pp, h_p, size, cudaMemcpyHostToDevice)); 
} 

int 
main(int argc, char **argv) 
{ 
    // init 
    // totally CPU codes   
    // ...... 
#if TIMING 
    double t1, t2, t3, t4, t5, t6; 
    t1 = get_time(); 
#endif 
#if CUDATIMING 
    cuda_timing_init; 
    cuda_timing_begin; 
#endif 
    // init data structure 
    // totally CPU codes 
    // ...... 
#if TIMING 
    t2 = get_time(); 
#endif 
#if CUDATIMING 
    cuda_timing_stop("init data structure"); 
    cuda_timing_begin; 
#endif 
    // establish context 
    cudaFree((void*)0); 
#if TIMING 
    t3 = get_time(); 
#endif 
#if CUDATIMING 
    cuda_timing_stop("establish context"); 
    cuda_timing_begin; 
#endif 
    // rearrange data 
    // totally CPU codes 
    // data on CPU side has different structure 
    // compared to data on GPU side, so I need 
    // to rearrange it. 
    // ...... 
#if TIMING 
    t4 = get_time(); 
#endif 
#if CUDATIMING 
    cuda_timing_stop("rearrange data"); 
    cuda_timing_begin; 
#endif 
    // copy data from host to device 
    // about 10 copies. the following are 2 of them 
     // all use copy_float/int_from_host_to_dev 
    // h_lap --> d_lap 
    copy_float_from_host_to_dev(h_lap, &d_lap, lapsize); 
    // h_etol --> d_etol 
    copy_int_from_host_to_dev(h_etol, &d_etol, etolsize); 
    // ...... 
#if TIMING 
    t5 = get_time(); 
#endif 
#if CUDATIMING 
    cuda_timing_stop("copy data"); 
    cuda_timing_begin; 
#endif 
    // time stepping 
    for(step = 1; step < para->nstep; step++) 
    { 
    /* kernel_1: matrix-vector multiplication. 
    * The matrix is special, so multiplication 
    * can be very fast. 
    * atomic operations are involved 
    * no data transfers between host and device */ 
    kernel_1<<<32768, 128>>>(......); 
    /* kernel_2: vector operations. 
    * Assuming that a,b,c,d are vectors, 
    * what kernel_2 does is: a=2*a-b+c*d 
    * no data transfers between host and device */ 
    kernel_2<<<16384, 128>>>(......); 
    } 
#if TIMING 
    t6 = get_time(); 
    printf("total time: %fs\n", t6-t1); 
    printf(" init data structure: %fs\n", t2-t1); 
    printf(" establish context: %fs\n", t3-t2); 
    printf(" rearrange data: %fs\n", t4-t3); 
    printf(" copy data: %fs\n", t5-t4); 
    printf(" time stepping: %fs\n", t6-t5); 
#endif 
#if CUDATIMING 
    cuda_timing_stop("time stepping"); 
    cuda_timing_destroy; 
#endif 

    // destroy data structure 
    // totally CPU codes 
    // ...... 

    return 0; 
} 

Answer 1

你只提供一個單一的代碼示例，所以只能提供一個答案：

花費的「建立上下文中，」由cudaEvent測量****（）的時間（0.0072 ms）與clock_gettime（）（〜20.5s）的測量值完全不同。實際上，這部分只有一條線建立了一個上下文。 cudaFree（0）這種巨大的差異是如何發生的？

你認爲cudaFree調用建立CUDA上下文不正確。懶惰的上下文建立發生在需要直接與上下文交互的第一個調用中。在這種情況下，您的事件計時代碼正在建立上下文，因此cudaFree調用基本上是免費的。這就是爲什麼兩種計時方法之間存在較大的掛鐘時間差。