2 GPUs के साथ cudaMalloc को कॉल करते समय खराब प्रदर्शन

Question

मेरे पास एक ऐसा एप्लिकेशन है जहां मैंने उपयोगकर्ता के सिस्टम पर GPUs के बीच प्रोसेसिंग लोड को विभाजित किया। असल में, प्रति GPU प्रति CPU थ्रेड है जो मुख्य अनुप्रयोग थ्रेड द्वारा समय-समय पर ट्रिगर होने पर GPU प्रसंस्करण अंतराल शुरू करता है।

GPU प्रसंस्करण अंतराल के उदाहरण के लिए निम्न छवि (एनवीआईडीआईए के CUDA प्रोफाइलर टूल का उपयोग करके उत्पन्न) पर विचार करें - यहां एप्लिकेशन एक एकल GPU का उपयोग कर रहा है।

आप देख सकते हैं GPU प्रसंस्करण समय के एक बड़े हिस्से दो छँटाई संचालन द्वारा उपयोग किया जाता है और मैं इस (जोर :: sort_by_key) के लिए थ्रस्ट लाइब्रेरी का उपयोग कर रहा हूँ। साथ ही, यह जोर देता है जैसे :: sort_by_key वास्तविक प्रकार शुरू करने से पहले हुड के नीचे कुछ cudaMallocs को कॉल करता है।

अब एक ही प्रसंस्करण अंतराल जहां आवेदन दो GPUs से अधिक संसाधन लोड फैल गया है पर विचार करें:

एक आदर्श दुनिया में आप 2 GPU प्रसंस्करण अंतराल बिल्कुल आधे की है कि होने की अपेक्षा करेंगे एकल जीपीयू (क्योंकि प्रत्येक जीपीयू आधा काम कर रहा है)। जैसा कि आप देख सकते हैं, यह मामला आंशिक रूप से नहीं है क्योंकि कुछ प्रकार के विवाद मुद्दे के कारण cudaMallocs को एक साथ (कभी-कभी 2-3 गुना अधिक) कहा जाता है। मुझे नहीं पता कि यह मामला क्यों होना चाहिए क्योंकि 2 जीपीयू के लिए मेमोरी आवंटन स्थान पूरी तरह से स्वतंत्र है इसलिए cudaMalloc पर सिस्टम-व्यापी लॉक नहीं होना चाहिए - एक प्रति-जीपीयू लॉक अधिक उचित होगा।

मेरी परिकल्पना साबित करने के लिए कि इस मुद्दे को एक साथ cudaMalloc कॉल के साथ है, मैं दो सीपीयू धागे (प्रत्येक GPU के लिए) प्रत्येक बुला cudaMalloc कई बार के साथ एक हास्यास्पद साधारण प्रोग्राम बनाया। मैं पहली बार इस कार्यक्रम के भाग गया ताकि अलग धागे एक ही समय में cudaMalloc कॉल नहीं करते: ~

आप इसे लेता है देखने के आवंटन प्रति 175 माइक्रोसेकंड। इसके बाद, मैं धागे एक साथ cudaMalloc बुला के साथ कार्यक्रम भागा:

यहाँ, प्रत्येक कॉल ~ 538 माइक्रोसेकंड या पिछले मामले से अधिक समय के लिए 3 बार ले लिया! कहने की जरूरत नहीं है, यह मेरे आवेदन को काफी धीमा कर रहा है और यह कारण है कि इस मुद्दे को 2 जीपीयू से अधिक खराब हो जाएगा।

मैंने लिनक्स और विंडोज पर इस व्यवहार को देखा है। लिनक्स पर, मैं एनवीडिया ड्राइवर संस्करण 319.60 का उपयोग कर रहा हूं और विंडोज़ पर मैं 327.23 संस्करण का उपयोग कर रहा हूं। मैं CUDA टूलकिट 5.5 का उपयोग कर रहा हूँ।

संभावित कारण: मैं इन परीक्षणों में जीटीएक्स 6 9 0 का उपयोग कर रहा हूं। यह कार्ड मूल रूप से एक ही इकाई में 2 जी 80-जैसी जीपीयू है। यह एकमात्र "बहु-जीपीयू" सेटअप है जिसे मैंने चलाया है, तो शायद 6 9 0 के 2 जीपीयू के बीच कुछ हार्डवेयर निर्भरता के साथ cudaMalloc समस्या का कुछ संबंध है?

Answer 1

समस्या संक्षेप में प्रस्तुत करना है और एक एक संभव समाधान दे:

cudaMalloc विवाद शायद चालक स्तर विवाद (संभवतः talonmies suggestsed डिवाइस संदर्भों स्विच करने के लिए जरूरत की वजह से) की वजह से उपजी है और एक में इस अतिरिक्त विलंबता से बच सकते हैं cudaMalloc-ing और अस्थायी बफर द्वारा पहले से महत्वपूर्ण प्रदर्शन खंड।

ऐसा लगता है कि मुझे शायद अपने कोड को दोबारा करने की आवश्यकता है ताकि मैं किसी भी सॉर्टिंग दिनचर्या को कॉल नहीं कर रहा हूं जो हुडा के नीचे cudaMalloc को कॉल करता है (मेरे मामले में जोर :: sort_by_key)। CUB library इस संबंध में आशाजनक लग रहा है। बोनस के रूप में, सीयूबी उपयोगकर्ता को एक सीयूडीए स्ट्रीम पैरामीटर भी दिखाता है, जो प्रदर्शन को बढ़ावा देने के लिए भी काम कर सकता है।

जोर से सीयूबी पर जाने के बारे में कुछ विवरणों के लिए CUB (CUDA UnBound) equivalent of thrust::gather देखें।

अद्यतन:

मैं शावक :: DeviceRadixSort :: SortPairs के पक्ष में जोर :: sort_by_key के लिए कॉल पीछे हट गई।
मेरे प्रति-अंतराल प्रसंस्करण समय से इस मुंडा मिलीसेकंड को करना। इसके अलावा बहु-जीपीयू विवाद समस्या ने स्वयं को हल कर लिया है - 2 जीपीयू तक ऑफलोडिंग प्रसंस्करण समय लगभग 50% तक गिर जाती है, जैसा कि अपेक्षित है।

Answer 2

मैं इसे अस्वीकरण के साथ पेश करूंगा: मैं एनवीआईडीआईए चालक के आंतरिक लोगों के लिए गुप्त नहीं हूं, इसलिए यह कुछ हद तक सट्टा है।

जितना धीमा आप देख रहे हैं वह एक ही डिवाइस के मॉलोक को कॉल करने वाले एकाधिक धागे से प्रतिस्पर्धा के कारण ड्राइवर स्तर की विवाद है। डिवाइस मेमोरी आवंटन के लिए कई ओएस सिस्टम कॉल की आवश्यकता होती है, जैसे ड्राइवर स्तर संदर्भ स्विचिंग करता है। दोनों परिचालनों में विलंबता की एक गैर-तुच्छ राशि है। यह संभव है कि जब आप दो धागे एक साथ मेमोरी आवंटित करते हैं और आवंटित करते हैं तो अतिरिक्त समय होता है, अतिरिक्त डिवाइस विलंबता दोनों उपकरणों पर स्मृति आवंटित करने के लिए आवश्यक सिस्टम कॉल के अनुक्रम के दौरान एक डिवाइस से दूसरे डिवाइस पर स्विच करने से होती है।

मैं कुछ तरीके के बारे में सोच सकते हैं आप कम करने के लिए यह सक्षम होना चाहिए: के लिए अपने स्वयं के कस्टम जोर स्मृति allocator लिख कर

• आप शून्य करने के लिए जोर स्मृति आवंटन का सिस्टम कॉल भूमि के ऊपर कम कर सकता है उपकरण जो प्रारंभिकरण के दौरान आवंटित स्मृति के स्लैब से काम करता है। यह प्रत्येक sort_by_key के भीतर सभी सिस्टम कॉल ओवरहेड से छुटकारा पायेगा, लेकिन आपके उपयोगकर्ता को लिखने का प्रयास मेमोरी मैनेजर गैर तुच्छ है। दूसरी तरफ यह आपके जोर कोड के बाकी को बरकरार रखता है।
• आप वैकल्पिक सॉफ़्टवेयर लाइब्रेरी पर स्विच कर सकते हैं और वापस अस्थायी स्मृति के आवंटन को प्रबंधित कर सकते हैं। यदि आप प्रारंभिक चरण में सभी आवंटन करते हैं, तो मेमोरी आवंटन की लागत प्रत्येक थ्रेड के जीवन पर लगभग शून्य तक बढ़ाया जा सकता है।

आधारित रेखीय बीजगणित कोड मैं लिखा है बहु GPU CUBLAS में, मैं दोनों विचारों संयुक्त और एक स्टैंडअलोन उपयोगकर्ता अंतरिक्ष डिवाइस स्मृति प्रबंधक जो एक आवंटित डिवाइस स्मृति पूल एक बार बंद काम करता है लिखा था। मैंने पाया कि इंटरमीडिएट डिवाइस मेमोरी आवंटन की सभी ओवरहेड लागत को हटाने से एक उपयोगी गति उत्पन्न हुई। आपके उपयोग के मामले एक समान रणनीति से लाभ हो सकता है।