क्लैंग का उपयोग करते हुए ओपनसीएल को पीटीएक्स में संकलित करते समय अनसुलझा बाहरी?

Question

मैं this SO answer पर दिए गए निर्देशों लेकिन अनुसरण कर रहा हूँ जब मैं जिसके परिणामस्वरूप PTX फ़ाइल मैं clBuild

ptxas fatal : Unresolved extern function 'get_group_id' 

में अनुवर्ती त्रुटि मिलती है PTX फ़ाइल मैं हर OpenCL समारोह कॉल के लिए निम्नलिखित है में मैं

का उपयोग चलाने का प्रयास

.func (.param .b64 func_retval0) get_group_id 
(
     .param .b32 get_group_id_param_0 
) 
; 

उपरोक्त ओपनसीएल रनटाइम द्वारा बनाई गई पीटीएक्स फाइलों में मौजूद नहीं है जब मैं इसे सीएल फ़ाइल प्रदान करता हूं। इसके बजाय यह उचित विशेष रजिस्टर है।

fatal error: error in backend: Cannot cast between two non-generic address spaces 

उन निर्देशों अभी भी मान्य हैं:

these instructions निम्न त्रुटि के साथ PTX संकलन करने के लिए LLVM आईआर दौरान (एक अलग libclc पुस्तकालय के खिलाफ लिंक) मुझे एक विभाजन गलती देता है के बाद? क्या मुझे कुछ और करना चाहिए?

मैं libclc, बजना 3.7, और Nvidia ड्राइवर 352,39

Answer 1

के नवीनतम संस्करण का उपयोग कर रहा समस्या यह है कि LLVM एक OpenCL डिवाइस कोड पुस्तकालय प्रदान नहीं करता है। llvm हालांकि एक जीपीयू थ्रेड की आईडी प्राप्त करने के लिए अंतर्निहित प्रदान करता है। अब आपको क्लैंग के बिल्टिन का उपयोग करके get_global_id आदि के अपने इम्प्लांटेशन लिखना होगा और इसे nvptx लक्ष्य के साथ llvm बिटकोड में संकलित करना होगा। अपने आईआर को पीटीएक्स में कम करने से पहले आप अपने संकलित ओपनसीएल मॉड्यूल के साथ अपनी डिवाइस लाइब्रेरी को लिंक करने के लिए llvm-link का उपयोग करें और यही वह है।

एक उदाहरण है कि कैसे आप इस तरह के एक समारोह लिखते थे:

#define __ptx_mad(a,b,c) ((a)*(b)+(c)) 

__attribute__((always_inline)) unsigned int get_global_id(unsigned int dimindx) { 
    switch (dimindx) { 
    case 0: return __ptx_mad(__nvvm_read_ptx_sreg_ntid_x(), __nvvm_read_ptx_sreg_ctaid_x(), __nvvm_read_ptx_sreg_tid_x()); 
    case 1: return __ptx_mad(__nvvm_read_ptx_sreg_ntid_y(), __nvvm_read_ptx_sreg_ctaid_y(), __nvvm_read_ptx_sreg_tid_y()); 
    case 2: return __ptx_mad(__nvvm_read_ptx_sreg_ntid_z(), __nvvm_read_ptx_sreg_ctaid_z(), __nvvm_read_ptx_sreg_tid_z()); 
    default: return 0; 
    } 
}