CUDA कर्नेल

Question

में विभिन्न अनुभागों को समय देना मेरे पास एक CUDA कर्नेल है जो डिवाइस फ़ंक्शंस की एक श्रृंखला को कॉल करता है।

प्रत्येक डिवाइस फ़ंक्शन के लिए निष्पादन समय प्राप्त करने का सबसे अच्छा तरीका क्या है?

डिवाइस कार्यों में से किसी एक में कोड के किसी अनुभाग के लिए निष्पादन समय प्राप्त करने का सबसे अच्छा तरीका क्या है?

Answer 1

अपने स्वयं के कोड में, मैं सटीक समय प्राप्त करने के लिए clock() फ़ंक्शन का उपयोग करता हूं। सुविधा के लिए, मैं मैक्रो

enum { 
    tid_this = 0, 
    tid_that, 
    tid_count 
    }; 
__device__ float cuda_timers[ tid_count ]; 
#ifdef USETIMERS 
#define TIMER_TIC clock_t tic; if (threadIdx.x == 0) tic = clock(); 
#define TIMER_TOC(tid) clock_t toc = clock(); if (threadIdx.x == 0) atomicAdd(&cuda_timers[tid] , (toc > tic) ? (toc - tic) : (toc + (0xffffffff - tic))); 
#else 
#define TIMER_TIC 
#define TIMER_TOC(tid) 
#endif 

ये तो साधन करने के लिए डिवाइस कोड इस्तेमाल किया जा सकता है इस प्रकार है:

__global__ mykernel (...) { 

    /* Start the timer. */ 
    TIMER_TIC 

    /* Do stuff. */ 
    ... 

    /* Stop the timer and store the results to the "timer_this" counter. */ 
    TIMER_TOC(tid_this); 

    } 

फिर आप मेजबान कोड में cuda_timers पढ़ सकते हैं।

कुछ नोट:

• एक प्रति ब्लॉक आधार, यानी अगर आपने 100 एक ही कर्नेल को क्रियान्वित ब्लॉक है, उनकी हर समय का योग संगृहीत की जाएंगी टाइमर काम करते हैं।
• यह कहकर, टाइमर मानता है कि ज़ीरोथ थ्रेड सक्रिय है, इसलिए सुनिश्चित करें कि आप कोड के संभवतः अलग-अलग हिस्से में इन मैक्रोज़ को कॉल न करें।
• टाइमर घड़ी की टिकों की संख्या को गिनती है। मिलीसेकंड की संख्या प्राप्त करने के लिए, इसे अपने डिवाइस पर जीएचजेड की संख्या से विभाजित करें और 1000 से गुणा करें।
• टाइमर आपके कोड को थोड़ा धीमा कर सकते हैं, यही कारण है कि मैंने उन्हें #ifdef USETIMERS में लपेट लिया ताकि आप उन्हें बंद कर सकें आसानी से।
• हालांकि clock()clock_t प्रकार के पूर्णांक मान देता है, मैं संचित मानों को float के रूप में संग्रहीत करता हूं, अन्यथा मान कर्नेल के लिए चारों ओर लपेटेंगे जो कुछ सेकंड (सभी ब्लॉक पर संचित) से अधिक समय लेते हैं।
• घड़ी काउंटर के आसपास के मामले में चयन (toc > tic) ? (toc - tic) : (toc + (0xffffffff - tic))) आवश्यक है।

पीएस यह this question पर मेरे उत्तर की एक प्रति है, जो पूरे कर्नेल के लिए आवश्यक समय के बाद से कई बिंदु नहीं मिला था।