J/Ph8dZddlZddlZddlmZddlmZmZGddeZGdd eZ Gd d eZ e Z e j Z dd Z d ZdZdS)a Expose each GPU devices directly. This module implements a API that is like the "CUDA runtime" context manager for managing CUDA context stack and clean up. It relies on thread-local globals to separate the context stack management of each thread. Contexts are also shareable among threads. Only the main thread can destroy Contexts. Note: - This module must be imported by the main-thread. N)contextmanager)driverUSE_NV_BINDINGcLeZdZfdZdZdZdZdZedZ xZ S) _DeviceListc|dkr5tj}dt|D}||_|St t ||S)NlstcPg|]#}ttj|$S)_DeviceContextManagerr get_device).0devids Z/var/www/html/test/jupyter/venv/lib/python3.11/site-packages/numba/cuda/cudadrv/devices.py z+_DeviceList.__getattr__..s;000*&*;E*B*BCC000)rget_device_countranger superr __getattr__)selfattrnumdevgpus __class__s rrz_DeviceList.__getattr__sn 5==,..F00!&v000DDHK[$''33D999rc|j|S)zB Returns the context manager for device *devnum*. )r )rdevnums r __getitem__z_DeviceList.__getitem__$sxrcJdd|jDS)Nz, c,g|]}t|Sr )str)rds rrz'_DeviceList.__str__..+s333Q#a&&333r)joinr rs r__str__z_DeviceList.__str__*s%yy33$(333444rc*t|jSN)iterr r%s r__iter__z_DeviceList.__iter__-sDH~~rc*t|jSr()lenr r%s r__len__z_DeviceList.__len__0s48}}rctj5}|j}|||cdddS ddddS#1swxYwYdS)zFReturns the active device or None if there's no active device N)rget_active_contextr)racrs rcurrentz_DeviceList.current3s & ( ( $BYF!F| $ $ $ $ $ $ $ $! $ $ $ $ $ $ $ $ $ $ $ $ $ $ $ $ $ $sAAA) __name__ __module__ __qualname__rrr&r*r-propertyr1 __classcell__)rs@rrrs : : : : :   555$$X$$$$$rrc0eZdZdZdZdZdZdZdZdS)r aU Provides a context manager for executing in the context of the chosen device. The normal use of instances of this type is from ``numba.cuda.gpus``. For example, to execute on device 2:: with numba.cuda.gpus[2]: d_a = numba.cuda.to_device(a) to copy the array *a* onto device 2, referred to by *d_a*. c||_dSr()_device)rdevices r__init__z_DeviceContextManager.__init__Is  rc,t|j|Sr()getattrr9)ritems rrz!_DeviceContextManager.__getattr__Lst|T***rcNt|jjdSr()_runtimeget_or_create_contextr9idr%s r __enter__z_DeviceContextManager.__enter__Os &&t|77777rc\|jdSr()r9get_primary_contextpop)rexc_typeexc_valexc_tbs r__exit__z_DeviceContextManager.__exit__Rs( ((**..00000rc.d|S)Nzr%)formatr%s rr&z_DeviceContextManager.__str__Vs+222===rN) r2r3r4__doc__r;rrCrJr&r rrr r =si  +++888111>>>>>rr cXeZdZdZdZedZdZdZdZ dZ dZ d Z d Z d S) _RuntimezEmulate the CUDA runtime context management. It owns all Devices and Contexts. Keeps at most one Context per Device ct|_tj|_tj|_tj|_dSr() rr threadinglocal_tlscurrent_thread _mainthreadRLock_lockr%s rr;z_Runtime.__init__asFMM O%% %355_&& rc#JKtj5|}|d}|| dV||n#||wxYw ddddS#1swxYwYdS)aEnsure a CUDA context is available inside the context. On entrance, queries the CUDA driver for an active CUDA context and attaches it in TLS for subsequent calls so they do not need to query the CUDA driver again. On exit, detach the CUDA context from the TLS. This will allow us to pickup thirdparty activated CUDA context in any top-level Numba CUDA API. N)rr/_get_attached_contextrA_set_attached_context)roldctxnewctxs rensure_contextz_Runtime.ensure_contextns & ( ( 3 3//11F//55F  & &v . . . 3**62222**622222 3 3 3 3 3 3 3 3 3 3 3 3 3 3 3 3 3 3s)?BA0B0BBBBc|-|}|||S|Strt|}||S)zReturns the primary context and push+create it if needed for *devnum*. If *devnum* is None, use the active CUDA context (must be primary) or create a new one with ``devnum=0``. )rY_get_or_create_context_uncachedrint_activate_context_for)rr attached_ctxs rrAz_Runtime.get_or_create_contextsa >5577L#;;FCCC## %V--f55 5rcl|j5tj5}|s-|dcdddcdddS|j|j}tr)t|j }t|j }n|j j }|j j }||kr$d}t| |||cdddcdddS#1swxYwY ddddS#1swxYwYdS)zbSee also ``get_or_create_context(devnum)``. This version does not read the cache. rNz5Numba cannot operate on non-primary CUDA context {:x})rWrr/rarrrErr`handlecontext_handlevalue RuntimeErrorrLprepare_for_use)rrr0ctx ctx_handle ac_ctx_handlemsgs rr_z(_Runtime._get_or_create_context_uncachedsZ  *,, *55a88               )BI.BBDDC%@%(__ (+B,=(>(> %(Z%5 (*(9(? !]22 4*3::m+D+DEEE'')))'                                          s@D)D D) B+D7 D)D D)D D))D-0D-c|j5|j|}|}|}|||urt d||cdddS#1swxYwYdS)NzCannot switch CUDA-context.)rWrrErYrgpush)rrgpur\ cached_ctxs rraz_Runtime._activate_context_fors Z  )F#C,,..F3355J%*F*B*B"#@AAA KKMMM                  sA A55A9<A9c.t|jddS)Nattached_context)r=rSr%s rrYz_Runtime._get_attached_contextsty"4d;;;rc||j_dSr()rSrr)rris rrZz_Runtime._set_attached_contexts%( """rctj tjtj|jkr|dSdS)zqClear all contexts in the thread. Destroy the context if and only if we are in the main thread. N)rpop_active_contextrQrTrU_destroy_all_contextsr%s rresetz_Runtime.resets` '))5 '))5  # % %)9 9 9  & & ( ( ( ( ( : 9rcB|jD]}|dSr()rrw)rros rrvz_Runtime._destroy_all_contextss,9  C IIKKKK  rN)r2r3r4rMr;rr]rAr_rarYrZrwrvr rrrOrOZs ' ' '33^3&666 8   <<<))) ) ) )rrOc6t|S)z^Get the current device or use a device by device number, and return the CUDA context. )r@rA)rs r get_contextrzs  ) )& 1 11rcFtjfd}|S)z A decorator that ensures a CUDA context is available when *fn* is executed. Note: The function *fn* cannot switch CUDA-context. czt5|i|cdddS#1swxYwYdSr()r@r])argskwsfns r_require_cuda_contextz.require_context.._require_cuda_contexts  $ $ & & $ $2t#s## $ $ $ $ $ $ $ $ $ $ $ $ $ $ $ $ $ $s 044) functoolswraps)rrs` rrequire_contextrs; _R$$$$$ ! rc8tdS)zReset the CUDA subsystem for the current thread. In the main thread: This removes all CUDA contexts. Only use this at shutdown or for cleaning up between tests. In non-main threads: This clear the CUDA context stack only. N)r@rwr rrrwrws NNrr()rMrrQ contextlibrrrobjectrr rOr@rrzrrwr rrrs-  %%%%%%********&$&$&$&$&$&&$&$&$R>>>>>F>>>:tttttvtttn 8::}2222 ! ! !     r