σ \ K]c@@sddlmZmZddlmZmZmZmZedƒZ d„Z d„Z d„Z d„Z d„Zd „Zd „Zd „Zd „Zd „Zd„Zd„Zd„Zdefd„ƒYZd„Zdefd„ƒYZd„Zdefd„ƒYZdS(i(tprint_functiontabsolute_import(ttypestirttypingtmacrosThis is a stub.cO@s t‚dS(s OpenCL get_global_id() N(t _stub_error(targstkargs((s.lib/python2.7/site-packages/numba/roc/stubs.pyt get_global_idscO@s t‚dS(s OpenCL get_local_id() N(R(RR((s.lib/python2.7/site-packages/numba/roc/stubs.pyt get_local_idscO@s t‚dS(s" OpenCL get_global_size() N(R(RR((s.lib/python2.7/site-packages/numba/roc/stubs.pytget_global_sizescO@s t‚dS(s! OpenCL get_local_size() N(R(RR((s.lib/python2.7/site-packages/numba/roc/stubs.pytget_local_sizescO@s t‚dS(s OpenCL get_group_id() N(R(RR((s.lib/python2.7/site-packages/numba/roc/stubs.pyt get_group_id$scO@s t‚dS(s! OpenCL get_num_groups() N(R(RR((s.lib/python2.7/site-packages/numba/roc/stubs.pytget_num_groups+scO@s t‚dS(s OpenCL get_work_dim() N(R(RR((s.lib/python2.7/site-packages/numba/roc/stubs.pyt get_work_dim2scO@s t‚dS(s0 OpenCL barrier() Example: # workgroup barrier + local memory fence hsa.barrier(hsa.CLK_LOCAL_MEM_FENCE) # workgroup barrier + global memory fence hsa.barrier(hsa.CLK_GLOBAL_MEM_FENCE) # workgroup barrier + global memory fence hsa.barrier() N(R(RR((s.lib/python2.7/site-packages/numba/roc/stubs.pytbarrier9scO@s t‚dS(sΕ OpenCL mem_fence() Example: # local memory fence hsa.mem_fence(hsa.CLK_LOCAL_MEM_FENCE) # global memory fence hsa.mem_fence(hsa.CLK_GLOBAL_MEM_FENCE) N(R(RR((s.lib/python2.7/site-packages/numba/roc/stubs.pyt mem_fenceJs cC@s t‚dS(s HSAIL wavebarrier N(R(((s.lib/python2.7/site-packages/numba/roc/stubs.pyt wavebarrierXscC@s t‚dS(s- HSAIL activelanepermute_wavewidth_* N(R(tsrctlaneidtidentityt useidentity((s.lib/python2.7/site-packages/numba/roc/stubs.pytactivelanepermute_wavewidth_scC@s t‚dS(sG AMDGCN Data Share intrinsic forwards permute (push semantics) N(R(tsrc_lanet dest_lane((s.lib/python2.7/site-packages/numba/roc/stubs.pyt ds_permutefscC@s t‚dS(sH AMDGCN Data Share intrinsic backwards permute (pull semantics) N(R(RR((s.lib/python2.7/site-packages/numba/roc/stubs.pyt ds_bpermutemstStubcB@s,eZdZdZdZd„Zd„ZRS(skA stub object to represent special objects which is meaningless outside the context of HSA-python. scC@std|ƒ‚dS(Ns%s is not instantiable(tNotImplementedError(tcls((s.lib/python2.7/site-packages/numba/roc/stubs.pyt__new__{scC@s|jS(N(t _description_(tself((s.lib/python2.7/site-packages/numba/roc/stubs.pyt__repr__~s((t__name__t __module__t__doc__R t __slots__RR"(((s.lib/python2.7/site-packages/numba/roc/stubs.pyRts  cC@svt|ƒ}t|ƒ}d}tj||dƒ}tj|tjtj|ƒtjƒ}t j ||d||fƒS(Nshsail.smem.alloctCR( t_legalize_shapetlenRtArrayRt signaturetUniTupletintptAnyRt Intrinsic(tshapetdtypetndimtfnametrestypetsig((s.lib/python2.7/site-packages/numba/roc/stubs.pyt shared_array‚s   'tsharedcB@s8eZdZdZejdededddgƒZRS(sshared namespace ss shared.arraytcallabletargnamesR0R1( R#R$R%R RtMacroR6tTruetarray(((s.lib/python2.7/site-packages/numba/roc/stubs.pyR7‹scC@sHt|tƒr|St|tƒr)|fStdjt|ƒƒƒ‚dS(Nsinvalid type for shape; got {0}(t isinstancettupletintt TypeErrortformatttype(R0((s.lib/python2.7/site-packages/numba/roc/stubs.pyR(”s tatomiccB@s*eZdZdZdefd„ƒYZRS(satomic namespace staddcB@seZdZRS(sCadd(ary, idx, val) Perform atomic ary[idx] += val (R#R$R%(((s.lib/python2.7/site-packages/numba/roc/stubs.pyRD€s(R#R$R%R RRD(((s.lib/python2.7/site-packages/numba/roc/stubs.pyRCŸsN(t __future__RRtnumbaRRRRRRR R R R R RRRRRRRRtobjectRR6R7R(RC(((s.lib/python2.7/site-packages/numba/roc/stubs.pyts("