Coverage for cuda/core/_utils/cuda_utils.pyx: 95.28%

233 statements  

« prev     ^ index     » next       coverage.py v7.15.0, created at 2026-07-03 01:38 +0000

1# SPDX-FileCopyrightText: Copyright (c) 2024-2026 NVIDIA CORPORATION & AFFILIATES. All rights reserved. 

2# 

3# SPDX-License-Identifier: Apache-2.0 

4  

5import functools 

6from functools import partial 

7import multiprocessing 

8import platform 

9import warnings 

10from collections import namedtuple 

11from collections.abc import Sequence 

12from contextlib import ExitStack 

13from typing import Any, Callable 

14  

15from cuda.bindings import driver as driver, nvrtc as nvrtc, runtime as runtime 

16  

17# Module-level annotations that reference `driver`, `nvrtc`, and `runtime` so 

18# that stubgen-pyx keeps these imports in the generated `.pyi` (it would 

19# otherwise trim them as unused). These names are not assigned, so they only 

20# affect __annotations__ and have no runtime cost. 

21_keep_driver_in_stub: 'driver.CUresult' 

22_keep_nvrtc_in_stub: 'nvrtc.nvrtcResult' 

23_keep_runtime_in_stub: 'runtime.cudaError_t' 

24  

25from cuda.bindings.nvvm import nvvmError 

26from cuda.bindings.nvjitlink import nvJitLinkError 

27  

28from cpython.buffer cimport PyObject_GetBuffer, PyBuffer_Release, Py_buffer, PyBUF_SIMPLE 

29  

30from cuda.bindings cimport cynvrtc, cynvvm, cynvjitlink 

31  

32from cuda.core._utils.driver_cu_result_explanations import DRIVER_CU_RESULT_EXPLANATIONS 

33from cuda.core._utils.runtime_cuda_error_explanations import RUNTIME_CUDA_ERROR_EXPLANATIONS 

34  

35  

36class CUDAError(Exception): 

37 pass 

38  

39  

40class NVRTCError(CUDAError): 

41 pass 

42  

43  

44  

45ComputeCapability = namedtuple("ComputeCapability", ("major", "minor")) 

46  

47  

48def cast_to_3_tuple(label: str, cfg: int | tuple[int, ...]) -> tuple[int, int, int]: 

49 cfg_orig = cfg 28 tbfeubgevbwbxbybzb9 ! # $ % ' ( ) * + , - . / : ; = ? @ [ ] ^ _ ` { | AbBbCbDbEbFbzfAfBf,iidCf!gjdDfkdEfldFfmdGfnd6bHfIfJfKfLff g MfGbobHbNfIbJbKbLbMbd h i j k } ~ abbbpbqbrbcbNbObPbQbRbheSbdb7 e DcEcieod9c!cebpd@h!G#G$G-iKg%G.iFcqdrdsdtdudvdwdxdydzdAdBdCdDdEdFdGdHdIdJdKdLdMdNdOdPdQdRdjeke/i'G:iGcOfPf;iQfRf=iSfTfleme(eneoeUfVfWfXfYfZf0f1f2f#g3f4f$g5f6fpeqe)e7f8f9frese!fteue#f$f%f'f(fSd)f*f%g#c'g(g*e+e)g*g+g,g-gTd$c.g,eUd/g:g;g+fVdvewe,f-fxeye-e.f/f?i:f;fzeAe.eBeCe=f?f@f[f]f^f_f`f{f=g|f}f?g~fagDeEe/ebgcgdgFeGeegHeIefggghgigjgWdkglg@g%c[g]g:e;e^g_g`g{g|gXd'c}g=eYd~gahbhmgZdJeKengogLeMe?epgqg@irgsg[itgug]ivgwg^ixgyg_izgAg`iBgCg{iDgEg|i[hBGCG5D(G]h

50 if isinstance(cfg, int): 28 tbfeubgevbwbxbybzb9 ! # $ % ' ( ) * + , - . / : ; = ? @ [ ] ^ _ ` { | AbBbCbDbEbFbzfAfBf,iidCf!gjdDfkdEfldFfmdGfnd6bHfIfJfKfLff g MfGbobHbNfIbJbKbLbMbd h i j k } ~ abbbpbqbrbcbNbObPbQbRbheSbdb7 e DcEcieod9c!cebpd@h!G#G$G-iKg%G.iFcqdrdsdtdudvdwdxdydzdAdBdCdDdEdFdGdHdIdJdKdLdMdNdOdPdQdRdjeke/i'G:iGcOfPf;iQfRf=iSfTfleme(eneoeUfVfWfXfYfZf0f1f2f#g3f4f$g5f6fpeqe)e7f8f9frese!fteue#f$f%f'f(fSd)f*f%g#c'g(g*e+e)g*g+g,g-gTd$c.g,eUd/g:g;g+fVdvewe,f-fxeye-e.f/f?i:f;fzeAe.eBeCe=f?f@f[f]f^f_f`f{f=g|f}f?g~fagDeEe/ebgcgdgFeGeegHeIefggghgigjgWdkglg@g%c[g]g:e;e^g_g`g{g|gXd'c}g=eYd~gahbhmgZdJeKengogLeMe?epgqg@irgsg[itgug]ivgwg^ixgyg_izgAg`iBgCg{iDgEg|i[hBGCG5D(G]h

51 cfg = (cfg,) 28 tbfeubgevbwbxbybzb9 ! # $ % ' ( ) * + , - . / : ; = ? @ [ ] ^ _ ` { | AbBbCbDbEbFbzfAfBfCfDfEfFfGf6bHfIfJfKfLff g MfGbobHbNfIbJbKbLbMbd h i j k } ~ abbbpbqbrbcbNbObPbQbRbheSbdb7 e DcEcieod9c!cebpd@h!G#G$GKg%GqdrdsdtdudvdwdxdydzdAdBdCdDdEdFdGdHdIdJdKdLdMdNdOdPdQdRdjeke'GGcOfPfQfRfSfTflemeneoeUfVfWfXfYfZf0f1f2f3f4f5f6fpeqe7f8f9frese!fteue#f$f%f'f(fSd)f*f#c$c+fVdvewe,f-fxeye.f/f:f;fzeAeBeCe=f?f@f[f]f^f_f`f{f|f}f~fagDeEebgcgdgFeGeegHeIefggghgigjgWdkglg%c'cmgZdJeKengogLeMepgqgrgsgtgugvgwgxgygzgAgBgCgDgEg[hBG

52 else: 

53 common = "must be an int, or a tuple with up to 3 ints" 2,iid!gjdkdldmdnd@h-iKg.iFc/i:i;i=i(e#g$g)e%g#c'g(g*e+e)g*g+g,g-gTd$c.g,eUd/g:g;g-e?i.e=g?g/e@g%c[g]g:e;e^g_g`g{g|gXd'c}g=eYd~gahbh?e@i[i]i^i_i`i{i|i[hCG5D(G]h

54 if not isinstance(cfg, tuple): 2,iid!gjdkdldmdnd@h-iKg.iFc/i:i;i=i(e#g$g)e%g#c'g(g*e+e)g*g+g,g-gTd$c.g,eUd/g:g;g-e?i.e=g?g/e@g%c[g]g:e;e^g_g`g{g|gXd'c}g=eYd~gahbh?e@i[i]i^i_i`i{i|i[hCG5D(G]h

55 raise ValueError(f"{label} {common} (got {type(cfg)})") 2(G

56 if len(cfg) > 3: 2,iid!gjdkdldmdnd@h-iKg.iFc/i:i;i=i(e#g$g)e%g#c'g(g*e+e)g*g+g,g-gTd$c.g,eUd/g:g;g-e?i.e=g?g/e@g%c[g]g:e;e^g_g`g{g|gXd'c}g=eYd~gahbh?e@i[i]i^i_i`i{i|i[hCG5D]h

57 raise ValueError(f"{label} {common} (got tuple with length {len(cfg)})") 2CG

58 if any(not isinstance(val, int) for val in cfg): 2,iid!gjdkdldmdnd@h-iKg.iFc/i:i;i=i(e#g$g)e%g#c'g(g*e+e)g*g+g,g-gTd$c.g,eUd/g:g;g-e?i.e=g?g/e@g%c[g]g:e;e^g_g`g{g|gXd'c}g=eYd~gahbh?e@i[i]i^i_i`i{i|i[h5D]h

59 raise ValueError(f"{label} {common} (got {cfg})") 25D

60 if any(val < 1 for val in cfg): 28 tbfeubgevbwbxbybzb9 ! # $ % ' ( ) * + , - . / : ; = ? @ [ ] ^ _ ` { | AbBbCbDbEbFbzfAfBf,iidCf!gjdDfkdEfldFfmdGfnd6bHfIfJfKfLff g MfGbobHbNfIbJbKbLbMbd h i j k } ~ abbbpbqbrbcbNbObPbQbRbheSbdb7 e DcEcieod9c!cebpd@h!G#G$G-iKg%G.iFcqdrdsdtdudvdwdxdydzdAdBdCdDdEdFdGdHdIdJdKdLdMdNdOdPdQdRdjeke/i'G:iGcOfPf;iQfRf=iSfTfleme(eneoeUfVfWfXfYfZf0f1f2f#g3f4f$g5f6fpeqe)e7f8f9frese!fteue#f$f%f'f(fSd)f*f%g#c'g(g*e+e)g*g+g,g-gTd$c.g,eUd/g:g;g+fVdvewe,f-fxeye-e.f/f?i:f;fzeAe.eBeCe=f?f@f[f]f^f_f`f{f=g|f}f?g~fagDeEe/ebgcgdgFeGeegHeIefggghgigjgWdkglg@g%c[g]g:e;e^g_g`g{g|gXd'c}g=eYd~gahbhmgZdJeKengogLeMe?epgqg@irgsg[itgug]ivgwg^ixgyg_izgAg`iBgCg{iDgEg|i[hBG]h

61 plural_s = "" if len(cfg) == 1 else "s" 2KgBG]h

62 raise ValueError(f"{label} value{plural_s} must be >= 1 (got {cfg_orig})") 2KgBG]h

63 return cfg + (1,) * (3 - len(cfg)) 28 tbfeubgevbwbxbybzb9 ! # $ % ' ( ) * + , - . / : ; = ? @ [ ] ^ _ ` { | AbBbCbDbEbFbzfAfBf,iidCf!gjdDfkdEfldFfmdGfnd6bHfIfJfKfLff g MfGbobHbNfIbJbKbLbMbd h i j k } ~ abbbpbqbrbcbNbObPbQbRbheSbdb7 e DcEcieod9c!cebpd@h!G#G$G-iKg%G.iFcqdrdsdtdudvdwdxdydzdAdBdCdDdEdFdGdHdIdJdKdLdMdNdOdPdQdRdjeke/i'G:iGcOfPf;iQfRf=iSfTfleme(eneoeUfVfWfXfYfZf0f1f2f#g3f4f$g5f6fpeqe)e7f8f9frese!fteue#f$f%f'f(fSd)f*f%g#c'g(g*e+e)g*g+g,g-gTd$c.g,eUd/g:g;g+fVdvewe,f-fxeye-e.f/f?i:f;fzeAe.eBeCe=f?f@f[f]f^f_f`f{f=g|f}f?g~fagDeEe/ebgcgdgFeGeegHeIefggghgigjgWdkglg@g%c[g]g:e;e^g_g`g{g|gXd'c}g=eYd~gahbhmgZdJeKengogLeMe?epgqg@irgsg[itgug]ivgwg^ixgyg_izgAg`iBgCg{iDgEg|i[h

64  

65  

66cdef int HANDLE_RETURN(cydriver.CUresult err) except?-1 nogil: 

67 if err != cydriver.CUresult.CUDA_SUCCESS: 2a ^N)G_N@e`N[e{N]e|N8 }Ntb~NfeaO*GbOubcOgedO+GeOvbfOwbgOhOxbiOybjOzbkO9 lO! mO# nO$ oO% pO' qO( rO) sO* tO+ uO, vO- wO. xO/ yO: zO; AO= BO? CO@ DO[ EO] FO^ GO_ HO` IO{ JO| KOLOMONOOOPOQOROSOTOchUOVOWOXOYOZO0O1O2O3O4O5O6O7O8O9O!O#O$O%O'O(O)O*O+O,O-O^e.O/O:O;O_e=O`e?O{e@O|e[OAb]O}e^O~e_Oaf`O{O|O}O~OBbaPCbbPDbcPEbdPFbeP(cfP)cgP*chP+ciP,cjP-ckP.clP/cmPnPoPpPqPrPdhsPehtPfhuPvPwPxPyPzPAPzfBPAfCPDPEPFPBfGP,G-GHPIPJPKPLPidMP.G/G}i~iNPOPPPQPRPSPTPCfUPVP:GWP;GXPYPZP0P1P2P3P4P5P!gjd6P7P8P=G9P?G!Paj#Pbj$P%P'P(P)P*P+P,P-P.P/P:P;P=P?P@P[P]P^P_P`P{P|P}P~PaQDfbQ@G[GcQdQeQfQgQkdhQ]G^GcjdjiQjQkQlQmQnQoQEfpQqQrQ_GsQ`GtQuQvQwQxQyQzQAQBQCQDQldEQFQGQ{GHQ|GIQejJQfjKQLQMQNQOQPQQQRQSQFfTQUQVQ}GWQ~GXQYQZQ0Q1Q2Q3Q4Q5Q6Q7Qmd8Q9Q!QaH#QbH$Qgj%Qhj'Q(Q)Q*Q+Q,Q-Q.Q/QGf:Q;Q=QcH?QdH@Q[Q]Q^Q_Q`Q{Q|Q}Q~QaRndbRcRdReHeRfHfRijgRjjhRiRjRkRlRmRnRoRpRqRrRsRtRuRvRwRxRyRzRARBRCRDRERFRGRHRIRJRKRLRMRNRORPRQRRRSRTRURVRWRghXRYRZR0R1R2R3R4R5R6R7R8R9R!R#Rhh$R%R'R(R)R6D*R6b+R,R-Rih.Rbf/R:R;R=R?R@R[R]R^R_R`R{R|R}R~RaSbScSdSeSfSgSHfhSIfiSJfjSKfkScflSdfmSLfnSgHoShHpSiHqS:crS0dsSjHtSkHuSf vSg wS7DxSMfySGbzSobASHbBSNfCSefDSjhESIbFSJbGSKbHSLbISMbJSKSLSMSNSOSPSQSRSd SSkhTSNeUSlhVSOeWSPeXSQeYSffZSgf0Shf1Sh 2Si 3Sj 4Sk 5S} 6S~ 7Sab8Sbb9Spb!Sqb#Srb$Sif%Sjf'Skf(Slf)Smf*Snf+S,S-S.Scb/SNb:SOb;SPb=SQb?SRb@She[SSb]Sdb^SlHmHofpf_Smhnhohphqhrhshth`Suh{SnHLg|SoHqfpHqHRe}SrfsftfrHsHuftHuHMg~SaTvhwhxhvfvHwfwHxHyHzHAHBHCHDHEHFHGHSebTxNHHyNIH8DJH9DKHzNLHANMHBNNHCNOH!DPH#DQHRHSH7 THe UHyhcTzhdT^heTHcfTAhgTBhhT_hiT`hjTkjkTlTmTDNVHENWHXHYHZHFN0HnT1H2H3HGN4HoT5H6H7HHN8HpT9H!H#HIN$HqT%H'H(HJN)HrT*H+H,HFg-HsT.H/HKN:HLN;HMN=HGg?HNN@HON[H1d]H2d^HtT;c_H=c`HTbuTUbvT?c{H@c|HPN}HQN~HRNaISNbIVbwTWbxTXbyTYbzTZbAT0bBTNgCTDTETFTGTHTITJTOgTeUeKTLTMTNTOTPTQTRTSTTTPgUTVTWTXTYTZT0T1T2T3T4T5T6T7T8T9T!T#T$T%T'T(T)T*T+T,T-T.T/T:T;T=T?T@T[T]T^T_T`T{T|T}T~TaUbUcUdUeUfUgUhUiUjUkUlUmUnUoUpUqUrUsUtUuUvUwUxUyUzUAUBUCUDUEUFUGUHUIUJUKULUMUNUOUPUQURUSUTUUUVUWUXUYUZU0U1U2U3U4U5U6U7U8U9U!U#U$U%U'U(U)U*U+U,U-U.U/U:U;U=U?U@U[U]U^U_U`U{U|U}U~UaVbVcVdVeVfVgVhViVjVkVlVmVnVoVChpVqVrVDhQgRgEh$DDc%DsVcItVdIuVeIvVfIwVgIxVhIyViIzVEcAVjIBVkICVlIDVmIEV'DFV(DGV)DHVieIVnIoIpIqIrIJVKVLVMVNV{hsItIuIvIOVwIxIPVyIzI|h}h~haiDGQVRVEGFhAIFGGhodljHh9c*D!c+DSVGGIhHGJhIGKhTVJGLhUVKGMhBILGNhVVWVXVYVZV0VMG1VNG2VOG3VPG4VQG5VRG6VSG7V8V9V!V#V$V%V'V(Veb/b:bpd)V*V+VFc,VqdrdsdtdudvdwdxdydzdAdBdCdDdEdFdGdHdIdJdKdLdMdNdOdPdQdRdjeke;b=b?b@b[b]b^b_b`b{b|b}b~bacbcccdcecfcgchcicjckclcmcncocpcqcrc,Dsc-V.V/VSgtcucTg:VUgVgWgCIDI;VXgYgZg0gOhPh=V?Vfb@V[Vgb]V^V_Vhb`V2b{V|V}V~Vl aWbWm cWn dWo eWp fWq gWr hWs iWt jWu kWlWv mWw nWx oWy pWqWz rWA sWtWuWvWwWibxWyWjbzWAWBWkbCW3bDWEWFWGWB HWIWC JWD KWE LWF MWG NWH OWI PWJ QWK RWSWL TWM UWN VWO WWXWP YWQ ZW0W1W2W3Wlb4W5Wmb6W7W8Wnb9W4b!W#W$W%WR 'W(WS )WT *WU +WV ,WW -WX .WY /WZ :W0 ;W=W1 ?W2 @W3 [W4 ]W^W5 _W6 `W{W|W1gEIFIGI}W[c~WaXHIbXIIcXdXeXfXVe-D.DgXJIhXKIiX/DjX:DkX;DlX=DmXIcnXWeoX]cpXXeqXLIrXMIsXNItXOIuXPIvXQIwXRIxXSIyXTIzXUIAXVIBXWICXXIDXYIEXZIFX0IGX?DHX1IIX@DJX2IKX[DLX3IMX]DNX4IOX^DPX5IQX_DRX6ISX`DTX7IUX{DVX8IWX9IXX!IYX#IZX$I0X%I1X'I2X(I3X)I4X*I5X+I6X,I7X-I8X.I9X/I!X:I#X;I$XTG%Xmj'XUG(XVG)Xnj*XWG+XYe,X-X2g3gZe=I?I@I.X^c/X0e1e[I:X]I;X^I=X_I?X|D@X}D[XJcKcsbLc]XMc^X_Xvc`X~D{XaE3d|XbEcE}XdE~XeEaYfEGcbYgE1bhEcYiEdYjEeYkEfYlEgYmEhYnEiYoEjYpEkYqElYrEsEtEuENcOcPcvEwExEQcRcScojmYpjnYoYyEpYzEqYAErYBEsYCEtYDEuYEEvYFEwYGExYHEyYIEzYJEAYKEBYLECYMEDYNEEYOEFYPEGYQEHYREIYSEJYTEKYUELYVEWEXEMYYENYZEOY0EPY1EQY2ERY3ESY4ETY5EUY6EVY7EWY8EXY9EYY!EZY#E0Y$E1Y%E2Y'E3Y(E4Y)E5Y*E6Y+E,E7Y-E8Y.E9Y/E!Y:E#Y;E$Y=E%Y?E'Y@E(Y[E)Y]E*Y^E+Y_E,Y`E-Y{E.Y|E/Y}E~EHg:Y2e;Y3e=Y_c?Y4d@Y[Y`IOf]Y^Y_Y`Y{I|I}I{Y|Y}Y~YaZaFPfbZ~IbFcZaJdZeZbJcFcJfZdJQfgZhZiZjZeJfJgJkZlZmZnZoZdFRfpZhJeFqZiJrZsZSfjJkJtZuZvZwZTfxZlJyZzZmJlenJoJXGYGpJqJrJsJtJqjme(euJrjsjvJwJxJneoeUfVfWfXfYfAZZfBZyJzJCZDZEZFZ0fGZAJHZIZJZKZBJ1fLZMZCJDJNZOZPZQZRZfF2f#gSZEJgFTZUZVZWZFJ3fXZYZZZGJHJIJ0Z1Z2Z3Z4ZhF4f$g5ZJJiF6ZKJ7Z8Z9Z5fLJMJ!Z#Z$Z%Z6f'ZNJ(Z)ZOJZGpePJQJRJSJTJUJVJWJXJYJjFqe)eZJtjkF0J1J2J3J4J5J6J7J8J9J!J#J$J%J7f'J(J*Z+Z,Z-Z)J.Z/Z:Z8f;Z=Z*J+J?Z@Z[Z]Z9f^Z,J_Z`Z{Z|Z}Z~Za0b0c0d0e0f0lFremFnFoFpFqFrFsFtFuFsevFQhwFxFyF!fteue#f$f%f'f(fSd)f*f%g#c'g(g*e+e)g*g+g,g-gTd$c.g,eUd/g:g;g+f-J.Jg0h0i0j0k0Vdujvj/J:J;J=J?J@JzFveAFBFCFDFEFFFGFHFIFJFweKFRhLFMFNFl0,fm0[J]Jn0o0p0q0-fr0^Js0t0_J0Gxe`J{J|J}J~JaKbKcKdKeKfKOFye-egKwjPFhKiKjKu0v0kK.fw0x0y0z0lKmKnKA0B0C0D0E0QF/fF0oKRFG0pKH0I0:fqKrKJ0K0L0M0;fN0sKO0P0tKzeuKvK1G2GwKxKyKzKAKxjAe.eBKyjzjCKDKEKBeCe=f?f@f[f]fQ0^fR0FKGKS0T0U0V0_fW0HKX0Y0Z000IK`f1020JKKK3040506070SF{f=g80LKTF90!0#0$0MK|f%0'0(0NKOKPK)0*0+0,0-0UF}f?g.0QKVF/0RK:0;0=0~fSKTK?0@0[0]0ag^0UK_0`0VK3GDeWKXKYKZK0K1K2K3K4K5KWFEe/e6KAjXF7K8K9K!K#K$K%K'K(K)K*K+K,K-Kbg.K/K{0|0}0~0:Ka1b1c1cgd1e1;K=Kf1g1h1i1dgj1?Kk1l1m1n1o1p1q1r1s1t1u1v1YFFeZF0F1F2F3F4F5F6F7FGe8FSh9F!F#FegHeIefggghgigjgWdkglg@g%c[g]g:e;e^g_g`g{g|gXd'c}g=eYd~gahbhmg@K[Kw1x1y1z1A1ZdBjCj]K^K_K`K{K|K$FJe%F'F(F)F*F+F,F-F.F/FKe:FTh;F=F?FB1ngC1}K~KD1E1F1G1ogH1aLI1J1bL4GLecLdLeLfLgLhLiLjLkLlLmL@FMe?enLDj[FoLpLqLK1L1M14e5eTcEj]F^FN1rLpgO1P1Q1R1sLtLuLS1T1U1V1W1_FqgX1vL`F{FIg|F}FJgY1wLZ10111xLrg21314151yLzLAL61718191!1~Fsg#1BLaG$1CL%1'1(1DLtg)1*1+1,1ELFLGL-1.1/1:1;1bGug=1HLcG?1IL@1[1]1JLvg^1_1`1{1KLLLML|1}1~1a2b2dGwgc2NLeGd2OLe2f2g2PLxgh2i2j2k2QLRLSLl2m2n2o2p2fGygq2TLgGr2ULs2t2u2VLzgv2w2x2y2WLXLYLz2A2B2C2D2hGAgE2ZLiGF20LG2H2I21LBgJ2K2L2M22L3L4LN2O2P2Q2R2jGCgS25LkGlGT26LU2V2W27LDgX2Y2Z2028L9L!L1222324252mGEg62#LnGoG72$L8292Uc!2Vc#25GWc$2Xc%2'2(2)2*2+2,2-2.2/2:2;2=2?2@2[2]2^2_2`2{2|2}2~2a3b3c3d3e3f3g3h3i3j3k3l3m3n3c o3Ycp3pG5bqG5dq3r3s3t3u3v3w3x3y3z3A3B3C3D3FjGj%LrGsGtGwcE3uG'L(LvG6GF3G3H3I3ZcJ3K3L3M3N3O3P3Q3R3S3T3U3V3W3X3Y3Z303)L132333435363738393!3wG#3$3*L%3+L'3(3,L)3*3+3,3-3.3/3:3;3=3?3@3[3]3^3_3`3{3|3}3~3a4b4c4d4e4f4g4h4i4j4k4l4m4n4o4p4q4r4s4t4u4v4w4x4y4-Lz4.LA4/LB4:LC4;LD4=LE4?LF40cG4@LH4biI4HjJ4IjK4L4M4N4O4P4JjQ4[LR4]LS4^LT4_LU4KjV4W4X4Y4Z4`L04{L14|L24}L34~L44aM54bM64cM74dM84eM94fM!4gM#4hM$4iM%4jM'4kM(4lM)4mM*4Lj+4nM,4Mj-4oM.4pM/4qM:4rM;4sM=4tM?4@4uM[4vM]4Uh^4_4`4{4|4}4~4a5b5c5d5e5f5g5h5i5j5k5l5m5n5o5p5q5r5s5t5u5v5wMw5xMyMx5zMAMBMCMy5z5DMA5B5EMC5FMGMD5HMIME5JMF5KMLMMMNMG5OMH5I5PMJ5K5QMRMSMTML5M5N5UMO5P5VMQ5R5S5xfT5U5V5W5X5Y5Z505152535455565758595!5#5$5%5'5(5)5*5+5,5-5.5/5:5;5=5?5@5[5]5

68 return _check_driver_error(err) 2chghhhihkhNelh{h|h}h~hai1b5d

69 return 0 2a ^N)G_N@e`N[e{N]e|N8 }Ntb~NfeaO*GbOubcOgedO+GeOvbfOwbgOhOxbiOybjOzbkO9 lO! mO# nO$ oO% pO' qO( rO) sO* tO+ uO, vO- wO. xO/ yO: zO; AO= BO? CO@ DO[ EO] FO^ GO_ HO` IO{ JO| KOLOMONOOOPOQOROSOTOchUOVOWOXOYOZO0O1O2O3O4O5O6O7O8O9O!O#O$O%O'O(O)O*O+O,O-O^e.O/O:O;O_e=O`e?O{e@O|e[OAb]O}e^O~e_Oaf`O{O|O}O~OBbaPCbbPDbcPEbdPFbeP(cfP)cgP*chP+ciP,cjP-ckP.clP/cmPnPoPpPqPrPdhsPehtPfhuPvPwPxPyPzPAPzfBPAfCPDPEPFPBfGP,G-GHPIPJPKPLPidMP.G/G}i~iNPOPPPQPRPSPTPCfUPVP:GWP;GXPYPZP0P1P2P3P4P5P!gjd6P7P8P=G9P?G!Paj#Pbj$P%P'P(P)P*P+P,P-P.P/P:P;P=P?P@P[P]P^P_P`P{P|P}P~PaQDfbQ@G[GcQdQeQfQgQkdhQ]G^GcjdjiQjQkQlQmQnQoQEfpQqQrQ_GsQ`GtQuQvQwQxQyQzQAQBQCQDQldEQFQGQ{GHQ|GIQejJQfjKQLQMQNQOQPQQQRQSQFfTQUQVQ}GWQ~GXQYQZQ0Q1Q2Q3Q4Q5Q6Q7Qmd8Q9Q!QaH#QbH$Qgj%Qhj'Q(Q)Q*Q+Q,Q-Q.Q/QGf:Q;Q=QcH?QdH@Q[Q]Q^Q_Q`Q{Q|Q}Q~QaRndbRcRdReHeRfHfRijgRjjhRiRjRkRlRmRnRoRpRqRrRsRtRuRvRwRxRyRzRARBRCRDRERFRGRHRIRJRKRLRMRNRORPRQRRRSRTRURVRWRghXRYRZR0R1R2R3R4R5R6R7R8R9R!R#Rhh$R%R'R(R)R6D*R6b+R,R-Rih.Rbf/R:R;R=R?R@R[R]R^R_R`R{R|R}R~RaSbScSdSeSfSgSHfhSIfiSJfjSKfkScflSdfmSLfnSgHoShHpSiHqS:crS0dsSjHtSkHuSf vSg wS7DxSMfySGbzSobASHbBSNfCSefDSjhESIbFSJbGSKbHSLbISMbJSKSLSMSNSOSPSQSRSd SSkhTSNeUSlhVSOeWSPeXSQeYSffZSgf0Shf1Sh 2Si 3Sj 4Sk 5S} 6S~ 7Sab8Sbb9Spb!Sqb#Srb$Sif%Sjf'Skf(Slf)Smf*Snf+S,S-S.Scb/SNb:SOb;SPb=SQb?SRb@She[SSb]Sdb^SlHmHofpf_Smhnhohphqhrhshth`Suh{SnHLg|SoHqfpHqHRe}SrfsftfrHsHuftHuHMg~SaTvhwhxhvfvHwfwHxHyHzHAHBHCHDHEHFHGHSebTxNHHyNIH8DJH9DKHzNLHANMHBNNHCNOH!DPH#DQHRHSH7 THe UHyhcTzhdT^heTHcfTAhgTBhhT_hiT`hjTkjkTlTmTDNVHENWHXHYHZHFN0HnT1H2H3HGN4HoT5H6H7HHN8HpT9H!H#HIN$HqT%H'H(HJN)HrT*H+H,HFg-HsT.H/HKN:HLN;HMN=HGg?HNN@HON[H1d]H2d^HtT;c_H=c`HTbuTUbvT?c{H@c|HPN}HQN~HRNaISNbIVbwTWbxTXbyTYbzTZbAT0bBTNgCTDTETFTGTHTITJTOgTeUeKTLTMTNTOTPTQTRTSTTTPgUTVTWTXTYTZT0T1T2T3T4T5T6T7T8T9T!T#T$T%T'T(T)T*T+T,T-T.T/T:T;T=T?T@T[T]T^T_T`T{T|T}T~TaUbUcUdUeUfUgUhUiUjUkUlUmUnUoUpUqUrUsUtUuUvUwUxUyUzUAUBUCUDUEUFUGUHUIUJUKULUMUNUOUPUQURUSUTUUUVUWUXUYUZU0U1U2U3U4U5U6U7U8U9U!U#U$U%U'U(U)U*U+U,U-U.U/U:U;U=U?U@U[U]U^U_U`U{U|U}U~UaVbVcVdVeVfVgVhViVjVkVlVmVnVoVChpVqVrVDhQgRgEh$DDc%DsVcItVdIuVeIvVfIwVgIxVhIyViIzVEcAVjIBVkICVlIDVmIEV'DFV(DGV)DHVieIVnIoIpIqIrIJVKVLVMVNVsItIuIvIOVwIxIPVyIzIDGQVRVEGFhAIFGGhodljHh9c*D!c+DSVGGIhHGJhIGKhTVJGLhUVKGMhBILGNhVVWVXVYVZV0VMG1VNG2VOG3VPG4VQG5VRG6VSG7V8V9V!V#V$V%V'V(Veb/b:bpd)V*V+VFc,VqdrdsdtdudvdwdxdydzdAdBdCdDdEdFdGdHdIdJdKdLdMdNdOdPdQdRdjeke;b=b?b@b[b]b^b_b`b{b|b}b~bacbcccdcecfcgchcicjckclcmcncocpcqcrc,Dsc-V.V/VSgtcucTg:VUgVgWgCIDI;VXgYgZg0gOhPh=V?Vfb@V[Vgb]V^V_Vhb`V2b{V|V}V~Vl aWbWm cWn dWo eWp fWq gWr hWs iWt jWu kWlWv mWw nWx oWy pWqWz rWA sWtWuWvWwWibxWyWjbzWAWBWkbCW3bDWEWFWGWB HWIWC JWD KWE LWF MWG NWH OWI PWJ QWK RWSWL TWM UWN VWO WWXWP YWQ ZW0W1W2W3Wlb4W5Wmb6W7W8Wnb9W4b!W#W$W%WR 'W(WS )WT *WU +WV ,WW -WX .WY /WZ :W0 ;W=W1 ?W2 @W3 [W4 ]W^W5 _W6 `W{W|W1gEIFIGI}W[c~WaXHIbXIIcXdXeXfXVe-D.DgXJIhXKIiX/DjX:DkX;DlX=DmXIcnXWeoX]cpXXeqXLIrXMIsXNItXOIuXPIvXQIwXRIxXSIyXTIzXUIAXVIBXWICXXIDXYIEXZIFX0IGX?DHX1IIX@DJX2IKX[DLX3IMX]DNX4IOX^DPX5IQX_DRX6ISX`DTX7IUX{DVX8IWX9IXX!IYX#IZX$I0X%I1X'I2X(I3X)I4X*I5X+I6X,I7X-I8X.I9X/I!X:I#X;I$XTG%Xmj'XUG(XVG)Xnj*XWG+XYe,X-X2g3gZe=I?I@I.X^c/X0e1e[I:X]I;X^I=X_I?X|D@X}D[XJcKcsbLc]XMc^X_Xvc`X~D{XaE3d|XbEcE}XdE~XeEaYfEGcbYgE1bhEcYiEdYjEeYkEfYlEgYmEhYnEiYoEjYpEkYqElYrEsEtEuENcOcPcvEwExEQcRcScojmYpjnYoYyEpYzEqYAErYBEsYCEtYDEuYEEvYFEwYGExYHEyYIEzYJEAYKEBYLECYMEDYNEEYOEFYPEGYQEHYREIYSEJYTEKYUELYVEWEXEMYYENYZEOY0EPY1EQY2ERY3ESY4ETY5EUY6EVY7EWY8EXY9EYY!EZY#E0Y$E1Y%E2Y'E3Y(E4Y)E5Y*E6Y+E,E7Y-E8Y.E9Y/E!Y:E#Y;E$Y=E%Y?E'Y@E(Y[E)Y]E*Y^E+Y_E,Y`E-Y{E.Y|E/Y}E~EHg:Y2e;Y3e=Y_c?Y4d@Y[Y`IOf]Y^Y_Y`Y{I|I}I{Y|Y}Y~YaZaFPfbZ~IbFcZaJdZeZbJcFcJfZdJQfgZhZiZjZeJfJgJkZlZmZnZoZdFRfpZhJeFqZiJrZsZSfjJkJtZuZvZwZTfxZlJyZzZmJlenJoJXGYGpJqJrJsJtJqjme(euJrjsjvJwJxJneoeUfVfWfXfYfAZZfBZyJzJCZDZEZFZ0fGZAJHZIZJZKZBJ1fLZMZCJDJNZOZPZQZRZfF2f#gSZEJgFTZUZVZWZFJ3fXZYZZZGJHJIJ0Z1Z2Z3Z4ZhF4f$g5ZJJiF6ZKJ7Z8Z9Z5fLJMJ!Z#Z$Z%Z6f'ZNJ(Z)ZOJZGpePJQJRJSJTJUJVJWJXJYJjFqe)eZJtjkF0J1J2J3J4J5J6J7J8J9J!J#J$J%J7f'J(J*Z+Z,Z-Z)J.Z/Z:Z8f;Z=Z*J+J?Z@Z[Z]Z9f^Z,J_Z`Z{Z|Z}Z~Za0b0c0d0e0f0lFremFnFoFpFqFrFsFtFuFsevFQhwFxFyF!fteue#f$f%f'f(fSd)f*f%g#c'g(g*e+e)g*g+g,g-gTd$c.g,eUd/g:g;g+f-J.Jg0h0i0j0k0Vdujvj/J:J;J=J?J@JzFveAFBFCFDFEFFFGFHFIFJFweKFRhLFMFNFl0,fm0[J]Jn0o0p0q0-fr0^Js0t0_J0Gxe`J{J|J}J~JaKbKcKdKeKfKOFye-egKwjPFhKiKjKu0v0kK.fw0x0y0z0lKmKnKA0B0C0D0E0QF/fF0oKRFG0pKH0I0:fqKrKJ0K0L0M0;fN0sKO0P0tKzeuKvK1G2GwKxKyKzKAKxjAe.eBKyjzjCKDKEKBeCe=f?f@f[f]fQ0^fR0FKGKS0T0U0V0_fW0HKX0Y0Z000IK`f1020JKKK3040506070SF{f=g80LKTF90!0#0$0MK|f%0'0(0NKOKPK)0*0+0,0-0UF}f?g.0QKVF/0RK:0;0=0~fSKTK?0@0[0]0ag^0UK_0`0VK3GDeWKXKYKZK0K1K2K3K4K5KWFEe/e6KAjXF7K8K9K!K#K$K%K'K(K)K*K+K,K-Kbg.K/K{0|0}0~0:Ka1b1c1cgd1e1;K=Kf1g1h1i1dgj1?Kk1l1m1n1o1p1q1r1s1t1u1v1YFFeZF0F1F2F3F4F5F6F7FGe8FSh9F!F#FegHeIefggghgigjgWdkglg@g%c[g]g:e;e^g_g`g{g|gXd'c}g=eYd~gahbhmg@K[Kw1x1y1z1A1ZdBjCj]K^K_K`K{K|K$FJe%F'F(F)F*F+F,F-F.F/FKe:FTh;F=F?FB1ngC1}K~KD1E1F1G1ogH1aLI1J1bL4GLecLdLeLfLgLhLiLjLkLlLmL@FMe?enLDj[FoLpLqLK1L1M14e5eTcEj]F^FN1rLpgO1P1Q1R1sLtLuLS1T1U1V1W1_FqgX1vL`F{FIg|F}FJgY1wLZ10111xLrg21314151yLzLAL61718191!1~Fsg#1BLaG$1CL%1'1(1DLtg)1*1+1,1ELFLGL-1.1/1:1;1bGug=1HLcG?1IL@1[1]1JLvg^1_1`1{1KLLLML|1}1~1a2b2dGwgc2NLeGd2OLe2f2g2PLxgh2i2j2k2QLRLSLl2m2n2o2p2fGygq2TLgGr2ULs2t2u2VLzgv2w2x2y2WLXLYLz2A2B2C2D2hGAgE2ZLiGF20LG2H2I21LBgJ2K2L2M22L3L4LN2O2P2Q2R2jGCgS25LkGlGT26LU2V2W27LDgX2Y2Z2028L9L!L1222324252mGEg62#LnGoG72$L8292Uc!2Vc#25GWc$2Xc%2'2(2)2*2+2,2-2.2/2:2;2=2?2@2[2]2^2_2`2{2|2}2~2a3b3c3d3e3f3g3h3i3j3k3l3m3n3c o3Ycp3pG5bqG5dq3r3s3t3u3v3w3x3y3z3A3B3C3D3FjGj%LrGsGtGwcE3uG'L(LvG6GF3G3H3I3ZcJ3K3L3M3N3O3P3Q3R3S3T3U3V3W3X3Y3Z303)L132333435363738393!3wG#3$3*L%3+L'3(3,L)3*3+3,3-3.3/3:3;3=3?3@3[3]3^3_3`3{3|3}3~3a4b4c4d4e4f4g4h4i4j4k4l4m4n4o4p4q4r4s4t4u4v4w4x4y4-Lz4.LA4/LB4:LC4;LD4=LE4?LF40cG4@LH4biI4HjJ4IjK4L4M4N4O4P4JjQ4[LR4]LS4^LT4_LU4KjV4W4X4Y4Z4`L04{L14|L24}L34~L44aM54bM64cM74dM84eM94fM!4gM#4hM$4iM%4jM'4kM(4lM)4mM*4Lj+4nM,4Mj-4oM.4pM/4qM:4rM;4sM=4tM?4@4uM[4vM]4Uh^4_4`4{4|4}4~4a5b5c5d5e5f5g5h5i5j5k5l5m5n5o5p5q5r5s5t5u5v5wMw5xMyMx5zMAMBMCMy5z5DMA5B5EMC5FMGMD5HMIME5JMF5KMLMMMNMG5OMH5I5PMJ5K5QMRMSMTML5M5N5UMO5P5VMQ5R5S5xfT5U5V5W5X5Y5Z505152535455565758595!5#5$5%5'5(5)5*5+5,5-5.5/5:5;5=5?5@5[5]5

70  

71  

72cdef int _get_current_device_id() except? -1: 

73 """Return the current thread's bound CUdevice ordinal.""" 

74 cdef cydriver.CUdevice dev 

75 with nogil: 2-L.L/L:L;L=L?L0c@LbiHjIjJj[L]L^L_LKjbMhMiMkMlMmMLjnMMjoMpMqMrMsMtMuMvM

76 HANDLE_RETURN(cydriver.cuCtxGetDevice(&dev)) 2-L.L/L:L;L=L?L0c@LbiHjIjJj[L]L^L_LKjbMhMiMkMlMmMLjnMMjoMpMqMrMsMtMuMvM

77 return <int>dev 2-L.L/L:L;L=L?L0c@LbiHjIjJj[L]L^L_LKjbMhMiMkMlMmMLjnMMjoMpMqMrMsMtMuMvM

78  

79  

80cdef int HANDLE_RETURN_NVRTC(cynvrtc.nvrtcProgram prog, cynvrtc.nvrtcResult err) except?-1 nogil: 

81 """Handle NVRTC result codes, raising NVRTCError with program log on failure.""" 

82 if err == cynvrtc.nvrtcResult.NVRTC_SUCCESS: 28 tbfeubgevbwbxbybzb9 ! # $ % ' ( ) * + , - . / : ; = ? @ [ ] ^ _ ` { | AbBbCbDbEbFbzfAfBfidCfjdDfkdEfldFfmdGfnd6bHfIfJfKfLff g 7DMfGbobHbNfIbJbKbLbMbd h i j k } ~ abbbpbqbrbcbNbObPbQbRbheSbdb7 e DcEcielj*D+DebpdFcqdrdsdtdudvdwdxdydzdAdBdCdDdEdFdGdHdIdJdKdLdMdNdOdPdQdRdjekeb ;b=b?b@b[b]b^b_b`b{b|b}b~bacbcccdcecfcgchcicjckclcmcncocpcqcrc,Dscvc~DaE3dbEcEdEeEfEGcgE1bhEiEjEkElEmEnEoEpEqErEsEtEuENcOcPcvEwExEQcRcScojpjyEzEAEBECEDEEEFEGEHEIEJEKELEMENEOEPEQERESETEUEVEWEXEYEZE0E1E2E3E4E5E6E7E8E9E!E#E$E%E'E(E)E*E+E,E-E.E/E:E;E=E?E@E[E]E^E_E`E{E|E}E~EOfaFPfbFcFQfdFRfeFSfTfleqjmesjneoeUfVfWfXfYfZf0f1ffF2fgF3fhF4fiF5f6fpejFqekF7f8f9flFremFnFoFpFqFrFsFtFuFsevFQhwFxFyF!fteue#f$f%f'f(fSd)f*f#cTd$cUd+fVdzFveAFBFCFDFEFFFGFHFIFJFweKFRhLFMFNF,f-fxeOFyePF.fQF/fRF:f;fzexjAezjBeCe=f?f@f[f]f^f_f`fSF{fTF|fUF}fVF~fagDeWFEeXFbgcgdgYFFeZF0F1F2F3F4F5F6F7FGe8FSh9F!F#FegHeIefggghgigjgWdkglg%cXd'cYdmgZd$FJe%F'F(F)F*F+F,F-F.F/FKe:FTh;F=F?FngogLe@FMe[F]F^Fpg_Fqg`F{F|Frg~FsgaGtgbGugcGvgdGwgeGxgfGyggGzghGAgiGBgjGCgkGlGDgmGEgnGoGUcVcWcXc6d7d8d`c{c|c9d!d#d$d1c2c}c%d~c'd(d)d*dad+dbdcddded,d-d.d/d:d;d=d?d@dfdgdhdc YcrGsGtGwcuGvGZc[d]d^d_d`d{d|d}dWMXMYMZM~d0Maebe1M

83 return 0 28 tbfeubgevbwbxbybzb9 ! # $ % ' ( ) * + , - . / : ; = ? @ [ ] ^ _ ` { | AbBbCbDbEbFbzfAfBfidCfjdDfkdEfldFfmdGfnd6bHfIfJfKfLff g 7DMfGbobHbNfIbJbKbLbMbd h i j k } ~ abbbpbqbrbcbNbObPbQbRbheSbdb7 e DcEcielj*D+DebpdFcqdrdsdtdudvdwdxdydzdAdBdCdDdEdFdGdHdIdJdKdLdMdNdOdPdQdRdjekeb ;b=b?b@b[b]b^b_b`b{b|b}b~bacbcccdcecfcgchcicjckclcmcncocpcqcrc,Dscvc~DaE3dbEcEdEeEfEGcgE1bhEiEjEkElEmEnEoEpEqErEsEtEuENcOcPcvEwExEQcRcScojpjyEzEAEBECEDEEEFEGEHEIEJEKELEMENEOEPEQERESETEUEVEWEXEYEZE0E1E2E3E4E5E6E7E8E9E!E#E$E%E'E(E)E*E+E,E-E.E/E:E;E=E?E@E[E]E^E_E`E{E|E}E~EOfaFPfbFcFQfdFRfeFSfTfleqjmesjneoeUfVfWfXfYfZf0f1ffF2fgF3fhF4fiF5f6fpejFqekF7f8f9flFremFnFoFpFqFrFsFtFuFsevFQhwFxFyF!fteue#f$f%f'f(fSd)f*f#cTd$cUd+fVdzFveAFBFCFDFEFFFGFHFIFJFweKFRhLFMFNF,f-fxeOFyePF.fQF/fRF:f;fzexjAezjBeCe=f?f@f[f]f^f_f`fSF{fTF|fUF}fVF~fagDeWFEeXFbgcgdgYFFeZF0F1F2F3F4F5F6F7FGe8FSh9F!F#FegHeIefggghgigjgWdkglg%cXd'cYdmgZd$FJe%F'F(F)F*F+F,F-F.F/FKe:FTh;F=F?FngogLe@FMe[F]F^Fpg_Fqg`F{F|Frg~FsgaGtgbGugcGvgdGwgeGxgfGyggGzghGAgiGBgjGCgkGlGDgmGEgnGoGUcVcWcXc6d7d8d`c{c|c9d!d#d$d1c2c}c%d~c'd(d)d*dad+dbdcddded,d-d.d/d:d;d=d?d@dfdgdhdc YcrGsGtGwcuGvGZc[d]d^d_d`d{d|d}dWMXMYMZM~d0Maebe1M

84 with gil: 1c

85 _raise_nvrtc_error(prog, err) 1c

86  

87  

88cdef int _raise_nvrtc_error(cynvrtc.nvrtcProgram prog, cynvrtc.nvrtcResult err) except -1: 

89 """Build error message with program log and raise NVRTCError.""" 

90 cdef const char* err_str = cynvrtc.nvrtcGetErrorString(err) 1c

91 cdef size_t logsize = 0 1c

92 if prog != NULL: 1c

93 cynvrtc.nvrtcGetProgramLogSize(prog, &logsize) 1c

94 cdef bytes log_bytes 

95 cdef str log_str = "" 1c

96 if logsize > 1 and prog != NULL: 1c

97 log_bytes = b" " * logsize 1c

98 if cynvrtc.nvrtcGetProgramLog(prog, <char*>log_bytes) == cynvrtc.nvrtcResult.NVRTC_SUCCESS: 1c

99 log_str = log_bytes.decode("utf-8", errors="backslashreplace") 1c

100 err_msg = f"{err}: {err_str.decode()}" if err_str != NULL else f"NVRTC error {err}" 1c

101 if log_str: 1c

102 err_msg += f", compilation log:\n\n{log_str}" 1c

103 raise NVRTCError(err_msg) 1c

104  

105  

106cdef int HANDLE_RETURN_NVVM(cynvvm.nvvmProgram prog, cynvvm.nvvmResult err) except?-1 nogil: 

107 """Handle NVVM result codes, raising nvvmError with program log on failure.""" 

108 if err == cynvvm.nvvmResult.NVVM_SUCCESS: 2}F2M3MpG5bqG5d4M5M6M7M8M9M!M#M$M%M'M(M)M*MFjGj

109 return 0 2}F2M3MpG5bqG5d4M5M6M7M8M9M!M#M$M%M'M(M)M*MFjGj

110 with gil: 25b

111 _raise_nvvm_error(prog, err) 25b

112  

113  

114cdef int _raise_nvvm_error(cynvvm.nvvmProgram prog, cynvvm.nvvmResult err) except -1: 

115 """Raise nvvmError annotated with the program log.""" 

116 cdef size_t logsize = 0 25b

117 if prog != NULL: 25b

118 cynvvm.nvvmGetProgramLogSize(prog, &logsize) 25b

119 cdef bytes log_bytes 

120 cdef str log_str = "" 25b

121 if logsize > 1 and prog != NULL: 25b

122 log_bytes = b" " * logsize 25b

123 if cynvvm.nvvmGetProgramLog(prog, <char*>log_bytes) == cynvvm.nvvmResult.NVVM_SUCCESS: 25b

124 log_str = log_bytes.decode("utf-8", errors="backslashreplace") 25b

125 cdef object exc = nvvmError(err) 25b

126 if log_str: 25b

127 exc.args = (exc.args[0] + f"\nNVVM program log: {log_str}", *exc.args[1:]) 25b

128 raise exc 25b

129  

130  

131cdef int HANDLE_RETURN_NVJITLINK( 

132 cynvjitlink.nvJitLinkHandle handle, cynvjitlink.nvJitLinkResult err) except?-1 nogil: 

133 """Handle nvJitLink result codes, raising nvJitLinkError with error log on failure.""" 

134 if err == cynvjitlink.nvJitLinkResult.NVJITLINK_SUCCESS: 2b +M,M-M.M/M:M;M=M?M@M[M7G8G]M^M_M`M{M|M}M~MaNbNcNdNeNyffNgNhNiNjNxGkNlNmNnNoNpNqNrNsNtN

135 return 0 2b +M,M-M.M/M:M;M=M?M@M[M7G8G]M^M_M`M{M|M}M~MaNbNcNdNeNfNgNhNiNjNxGkNlNmNnNoNpNqNrNsNtN

136 with gil: 2b yf

137 _raise_nvjitlink_error(handle, err) 2b yf

138  

139  

140cdef int _raise_nvjitlink_error( 

141 cynvjitlink.nvJitLinkHandle handle, cynvjitlink.nvJitLinkResult err) except -1: 

142 """Raise nvJitLinkError annotated with the error log.""" 

143 cdef size_t logsize = 0 2b yf

144 if handle != NULL: 2b yf

145 cynvjitlink.nvJitLinkGetErrorLogSize(handle, &logsize) 1b

146 cdef bytes log_bytes 

147 cdef str log_str = "" 2b yf

148 if logsize > 1 and handle != NULL: 2b yf

149 log_bytes = b" " * logsize 1b

150 if cynvjitlink.nvJitLinkGetErrorLog(handle, <char*>log_bytes) == \ 1b

151 cynvjitlink.nvJitLinkResult.NVJITLINK_SUCCESS: 1b

152 log_str = log_bytes.decode("utf-8", errors="backslashreplace") 1b

153 cdef object exc = nvJitLinkError(err) 2b yf

154 if log_str: 2b yf

155 exc.args = (exc.args[0] + f"\nnvJitLink error log: {log_str}", *exc.args[1:]) 1b

156 raise exc 2b yf

157  

158  

159cdef object _RUNTIME_SUCCESS = runtime.cudaError_t.cudaSuccess 

160cdef object _NVRTC_SUCCESS = nvrtc.nvrtcResult.NVRTC_SUCCESS 

161  

162  

163cpdef inline int _check_driver_error(cydriver.CUresult error) except?-1 nogil: 

164 if error == cydriver.CUresult.CUDA_SUCCESS: 2a Nj@eOj[ePj]eQj8 RjtbSjTjUjubVjWjXjvbYjwbZj0jxb1jyb2jzb3j9 4j! 5j# 6j$ 7j% 8j' 9j( !j) #j* $j+ %j, 'j- (j. )j/ *j: +j; ,j= -j? .j@ /j[ :j] ;j^ =j_ ?j` @j{ [j| ]j^j_j`j{jch|j}j~jakbkckdkekfkgkhkikjkkklk^emknkok_epk`eqk{erk|eskAbtk}euk~evkafwkxkykBbzkCbAkDbBkEbCkFbDk(cEk)cFk*cGk+cHk,cIk-cJk.cKk/cLkMkNkOkPkQkRkSkdhTkehUkfhVkWkXkYkZk0k1k2k3k4k5k6k7k8k9k!k#k$k%k'k(k)k*k+k,k-k.k/k:k;k=k?k@k[k]k^k_k`k{k|k}k~kalblcldlelflglhliljlklllmlnlolplqlrlsltlulvlwlxlylzlAlBlClDlElFlGlHlIlJlKlLlMlNlOlPlQlRlSlTlUlVlWlXlYlZl0l1l2l3l4l5l6l7l8l9l!l#l$l%l'l(l)l*l+l,l-l.l/l:l;l=l?l@l[l]l^l_l`l{l|l}l~lambmcmdmemfmgmhmimjmkmlmmmnmompmqmrmsmtmumghvmwmxmymzmAmBmCmhhDmEmFmGm6bHmImihJmbfKmLmMmNmOmPmQmRmSmTmUmVmWmXmYmZmcf0mdf1m2m3m4m5m:c6m0d7m8m9mf !mg #m$m%mGb'mob(mHb)m*mef+mjh,mIb-mJb.mKb/mLb:mMb;m=m?m@m[md ]mkh^mNe_mlh`mOePeQeffgfhfh i j k } ~ abbbpbqbrbifjfkflfmfnf{mcb|mNb}mOb~mPbanQbbnRbcndnSbendbfngnofpfhninmhnhciohphdiqhrheishthfiuhgijnLgknqfhilnRemnrfsfiitfnnufjionMgpnvhkiwhxhlivfqnwfrnsntnunvnwnSexn7 e HcFgGg1d2d;c=cTbUb?c@cVbWbXbYbZb0bynznAnBnCnDnEnyGmiFnGnHnInJnKnLnMnNnOnPnQnRnSnChTnUnVnDhWnEhDcXnYnZn0n1n2n3n4nEc5n6n7n8n9n!n#n$n%n{h|h}h~hai'nFhGhodHh9c(n!c)nIhJhKh*nLhMhNh+n,n-n.n/n:n;n=n?n@n[n]n^n_n`n{n|n}n~neb/b:baobocodoqdrdsdtdudvdwdxdydzdAdBdCdDdEdFdGdHdIdJdKdLdMdNdOdPdQdRdjekeeofogohoiojokolomonooopoqorosotouovowoxoyozoAoBoCoDoEoFoGoHoIoJoKoLoMoNoSgtcucTgUgVgWgXgYgZg0gOhPhce3cfbxcgb4cychb2b6e7e8el %bm n o p q r s t u 'bv w x y 7bz A 8b(bde5cibzcjb6cAckb3b9e!e#eB )bC D E F G H I J K *bL M N O 9bP Q !b+bee7clbBcmb8cCcnb4b$e%e'eR ,bS T U V W X Y Z 0 -b1 2 3 4 #b5 6 $b.b1gOoPoQoVeRoSoToUoVoWoXoYoZo0o1o2o3o4o5o6o7o8o9o!o#o$o%o'o(o)o*o+o,o-o.o/o:o;o=o?o@o[o]o^o_o`o{o|o}o~oapbpcpdpepfpgphpipjpkplpmpnpopppqprpsp2g3gZetpupvpwpxpyp|Dzp}DApJcKcsbLcBpMcCpDpEp3dFpGpHpIpJpKpGcLpMp1bNpOpPpQpRpSpTpUpVpWpXpYpZp0p1p2p3p4p5p6p7p8p9p!p#p$p%p'p(p)p*p+p,p-p.p/p:p;p=p?p@p[p]p^p_p`p{p|p}p~paqbqcqdqeqfqgqhqiqjqkqlqmqnqoqpqqqrqsqtquqvqwqxqyqzqAqBqCqDqEqFqGqHqIqJqKqLqMqNqOqPqQqRqSqTqUqVqWqXqYqZq0q1q2q3q4q5q6q7q8q9q!q#q$q%q'q(q)q*q+q,q-q.q/q:q;q=q?q@q[q]q^q_q`q{q|q}q~qarbrcrdrerfrgrhrirjrkrlrmrnrorprqrrrsrtrurvrwrxryrzrArBrCrDrErFrGrHrIrJrKrLrMrNrOrPrQrRrSrTrUrVrWrXrYrZr0r1r2r3r4r5r6r7r8r9r!r#r$r%r'r(r)r*r+r,r-r.r/r:r;r=r?r@r[r]r^r_r`r{r|r}r~rasbscsdsesfsgshsisjskslsmsnsospsqsrssstsusvswsxsyszsAsBsCsDsEsFsGsHsIsJsKsLsMsNsOsPsQsRsSsTsUsVsWsXsYsZs0s1s2s3s4s5s6s7s8s9s!s#s$s%s's(s)s*s+s,s-s.s/s:s;s=s?s@s[s]s^s_s`s{s|s}s~satbtctdtetftgthtitjtktltmtntotptqtrtstttutvtwtxtytztAtBtCtDtEtFtGtHtItJtKtLtMtNtOtPtQtRtStTtUtVtWtXtYtZt0t1t2t3t4t5t6t7t8t9t!t#t$t%t't(t)t*t+t,t-t.t/t:t;t=t?t@t[t]t^t_t`t{t|t}t~taubucudueufuguhuiujukulumunuoupuqurusutuuuvuwuxuyuzuAuBuCuDuEuFuGuHuIuJuKuLuMuNuOuPuQuRuSuTuUuVuWuXuYuZu0u1u2u3u4u5u6u7u8u9u!u#u$u%u'u(u)u*u+u,u-u.u/u:u;u=u?u@u[u]u^u_u`u{u|u}u~uavbvcvdvevfvgvhvivjvkvlvmvnvovpvqvrvsvtvuvvvwvxvyvzvAvBvCvDvEvFvGvHvIvJvKvLvMvNvOvPvQvRvSvTvUvVvWvXvYvZv0v1v2v3v4v5v6v7v8v9v!v#v$v%v'v(v)v*v+v,v-v.v/v:v;v=v?v@v[v]v^v_v`v{v|v}v~vawbwcwdwewfwgwhwiwjwkwlwmwnwowpwqwrwswtwuwvwwwxwywzwAwBwCwDwEwFwGwHwIwJwKwLwMwNwOwPwQwRwSwTwUwVwWwXwYwZw0w1w2w3w4w5w6w7w8w9w!w#w$w%w'w(w)w*w+w,w-w.w/w:w;w=w?w@w[w]w^w_w`w{w|w}w~waxbxcxdxexfxgxhxixjxkxlxmxnxoxpxqxrxsxtxuxvxwxxxyxzxAxBxCxDxExFxGxHxIxJxKxLxMxNxOxPxQxRxSxTxUxVxWxXxYxZx0x1x2x3x4x5x6x7x8x9x!x#x$x%x'x(x)x*x+x,x-x.x/x:x;x=x?x@x[x]x^x_x`x{x|x}x~xaybycydyeyTcfygyhyiyjykylymynyoypyqyrysytyuyvywyxyyyzyAyByCyDyEyFyGyHyIyJyKyLyMyNyOyPyQyRySyTyUyVyWyXyYyZy0y1y2y3y4y5y6y7y8y9y!y#y$y%y'y(y)y*y+y,y-y.y/y:y;y=y?y@y[y]y^y_y`y{y|y}y~yazbzczdzezfzgzhzizjzkzlzmznzozpzqzrzsztzuzvzwzxzyzzzAzBzCzDzEzFzGzHzIzJzKzLzMzNzOzPzQzRzSzTzUzVzWzXzYzZz0z1z2z3z4z5z6z7z8z9z!z#z$z%z'z(z)z*z+z,z-z.z/z:z;z=z?z@z[z]z^z_z`z{z|z}z~zaAbAcAdAeAfAgAhAiAjAkAlAmAnAoApAqArAsAtAuAvAwAxAyAzAAABACADAEAFAGAHAIAJAKALAMANAOAPAQARASATAUAVAWAXAYAZA0A1A2A3A4A5A6A7A8A9A!A#A$A%A'A(A)A*A+A5d,A-A.A/A:A;A=A?A@A[A]A^A_A`A{A|A}A~AaBbBcBdBeBfBgBhBiBjBkBlBmBnBoBpBqBrBsBtBuBvBwBxByBzBABBBCBDBEBFBGBHBIBJBKBLBMBNBOBPBQBRBSBTBUBVBWBXBYBZB0B1B2B3B4B5B6B7B8B9B!B#B$B%B'B(B)B*B+B,B-B.B/B:B;B=B?B@B[B]B^B_B`B{B|B}B~BaCbCcC0cdCeCfCgChCiCjCkClCmCnCoCpCqCrCsCtCuCvCwCxCyCzCACBCCCDCECFCGCHCICJCKCLCMCNCOCPCQCRCSCTCUCVCWCXCYCZC0CUh1C2C3C4C5C6C7C8C9C!C#C$C%C'C(C)C*C+C,C-C.C/C:C;C=C?C@C[C]C^C_C`C{C|C}C~CaDbDcDdDeDfDgDhDiDjDkDlDmDnDoDpDqDrDsDtDuDvDwDxDxfyDzDADBDCDDDEDFDGDHDIDJDKDLDMDNDODPDQDRDSDTDUDVDWDXDYDZD0D1D2D

165 return 0 2a Nj@eOj[ePj]eQj8 RjtbSjTjUjubVjWjXjvbYjwbZj0jxb1jyb2jzb3j9 4j! 5j# 6j$ 7j% 8j' 9j( !j) #j* $j+ %j, 'j- (j. )j/ *j: +j; ,j= -j? .j@ /j[ :j] ;j^ =j_ ?j` @j{ [j| ]j^j_j`j{j|j}j~jakbkckdkekfkgkhkikjkkklk^emknkok_epk`eqk{erk|eskAbtk}euk~evkafwkxkykBbzkCbAkDbBkEbCkFbDk(cEk)cFk*cGk+cHk,cIk-cJk.cKk/cLkMkNkOkPkQkRkSkdhTkehUkfhVkWkXkYkZk0k1k2k3k4k5k6k7k8k9k!k#k$k%k'k(k)k*k+k,k-k.k/k:k;k=k?k@k[k]k^k_k`k{k|k}k~kalblcldlelflglhliljlklllmlnlolplqlrlsltlulvlwlxlylzlAlBlClDlElFlGlHlIlJlKlLlMlNlOlPlQlRlSlTlUlVlWlXlYlZl0l1l2l3l4l5l6l7l8l9l!l#l$l%l'l(l)l*l+l,l-l.l/l:l;l=l?l@l[l]l^l_l`l{l|l}l~lambmcmdmemfmgmhmimjmkmlmmmnmompmqmrmsmtmumvmwmxmymzmAmBmCmDmEmFmGm6bHmImJmbfKmLmMmNmOmPmQmRmSmTmUmVmWmXmYmZmcf0mdf1m2m3m4m5m:c6m0d7m8m9mf !mg #m$m%mGb'mob(mHb)m*mef+mjh,mIb-mJb.mKb/mLb:mMb;m=m?m@m[md ]m^mNe_m`mOePeQeffgfhfh i j k } ~ abbbpbqbrbifjfkflfmfnf{mcb|mNb}mOb~mPbanQbbnRbcndnSbendbfngnofpfhninmhnhciohphdiqhrheishthfiuhgijnLgknqfhilnRemnrfsfiitfnnufjionMgpnvhkiwhxhlivfqnwfrnsntnunvnwnSexn7 e HcFgGg1d2d;c=cTbUb?c@cVbWbXbYbZb0bynznAnBnCnDnEnmiFnGnHnInJnKnLnMnNnOnPnQnRnSnChTnUnVnDhWnEhDcXnYnZn0n1n2n3n4nEc5n6n7n8n9n!n#n$n%n'nFhGhodHh9c(n!c)nIhJhKh*nLhMhNh+n,n-n.n/n:n;n=n?n@n[n]n^n_n`n{n|n}n~neb/b:baobocodoqdrdsdtdudvdwdxdydzdAdBdCdDdEdFdGdHdIdJdKdLdMdNdOdPdQdRdjekeeofogohoiojokolomonooopoqorosotouovowoxoyozoAoBoCoDoEoFoGoHoIoJoKoLoMoNoSgtcucTgUgVgWgXgYgZg0gOhPhce3cfbxcgb4cychb2b6e7e8el %bm n o p q r s t u 'bv w x y 7bz A 8b(bde5cibzcjb6cAckb3b9e!e#eB )bC D E F G H I J K *bL M N O 9bP Q !b+bee7clbBcmb8cCcnb4b$e%e'eR ,bS T U V W X Y Z 0 -b1 2 3 4 #b5 6 $b.b1gOoPoQoVeRoSoToUoVoWoXoYoZo0o1o2o3o4o5o6o7o8o9o!o#o$o%o'o(o)o*o+o,o-o.o/o:o;o=o?o@o[o]o^o_o`o{o|o}o~oapbpcpdpepfpgphpipjpkplpmpnpopppqprpsp2g3gZetpupvpwpxpyp|Dzp}DApJcKcsbLcBpMcCpDpEp3dFpGpHpIpJpKpGcLpMp1bNpOpPpQpRpSpTpUpVpWpXpYpZp0p1p2p3p4p5p6p7p8p9p!p#p$p%p'p(p)p*p+p,p-p.p/p:p;p=p?p@p[p]p^p_p`p{p|p}p~paqbqcqdqeqfqgqhqiqjqkqlqmqnqoqpqqqrqsqtquqvqwqxqyqzqAqBqCqDqEqFqGqHqIqJqKqLqMqNqOqPqQqRqSqTqUqVqWqXqYqZq0q1q2q3q4q5q6q7q8q9q!q#q$q%q'q(q)q*q+q,q-q.q/q:q;q=q?q@q[q]q^q_q`q{q|q}q~qarbrcrdrerfrgrhrirjrkrlrmrnrorprqrrrsrtrurvrwrxryrzrArBrCrDrErFrGrHrIrJrKrLrMrNrOrPrQrRrSrTrUrVrWrXrYrZr0r1r2r3r4r5r6r7r8r9r!r#r$r%r'r(r)r*r+r,r-r.r/r:r;r=r?r@r[r]r^r_r`r{r|r}r~rasbscsdsesfsgshsisjskslsmsnsospsqsrssstsusvswsxsyszsAsBsCsDsEsFsGsHsIsJsKsLsMsNsOsPsQsRsSsTsUsVsWsXsYsZs0s1s2s3s4s5s6s7s8s9s!s#s$s%s's(s)s*s+s,s-s.s/s:s;s=s?s@s[s]s^s_s`s{s|s}s~satbtctdtetftgthtitjtktltmtntotptqtrtstttutvtwtxtytztAtBtCtDtEtFtGtHtItJtKtLtMtNtOtPtQtRtStTtUtVtWtXtYtZt0t1t2t3t4t5t6t7t8t9t!t#t$t%t't(t)t*t+t,t-t.t/t:t;t=t?t@t[t]t^t_t`t{t|t}t~taubucudueufuguhuiujukulumunuoupuqurusutuuuvuwuxuyuzuAuBuCuDuEuFuGuHuIuJuKuLuMuNuOuPuQuRuSuTuUuVuWuXuYuZu0u1u2u3u4u5u6u7u8u9u!u#u$u%u'u(u)u*u+u,u-u.u/u:u;u=u?u@u[u]u^u_u`u{u|u}u~uavbvcvdvevfvgvhvivjvkvlvmvnvovpvqvrvsvtvuvvvwvxvyvzvAvBvCvDvEvFvGvHvIvJvKvLvMvNvOvPvQvRvSvTvUvVvWvXvYvZv0v1v2v3v4v5v6v7v8v9v!v#v$v%v'v(v)v*v+v,v-v.v/v:v;v=v?v@v[v]v^v_v`v{v|v}v~vawbwcwdwewfwgwhwiwjwkwlwmwnwowpwqwrwswtwuwvwwwxwywzwAwBwCwDwEwFwGwHwIwJwKwLwMwNwOwPwQwRwSwTwUwVwWwXwYwZw0w1w2w3w4w5w6w7w8w9w!w#w$w%w'w(w)w*w+w,w-w.w/w:w;w=w?w@w[w]w^w_w`w{w|w}w~waxbxcxdxexfxgxhxixjxkxlxmxnxoxpxqxrxsxtxuxvxwxxxyxzxAxBxCxDxExFxGxHxIxJxKxLxMxNxOxPxQxRxSxTxUxVxWxXxYxZx0x1x2x3x4x5x6x7x8x9x!x#x$x%x'x(x)x*x+x,x-x.x/x:x;x=x?x@x[x]x^x_x`x{x|x}x~xaybycydyeyTcfygyhyiyjykylymynyoypyqyrysytyuyvywyxyyyzyAyByCyDyEyFyGyHyIyJyKyLyMyNyOyPyQyRySyTyUyVyWyXyYyZy0y1y2y3y4y5y6y7y8y9y!y#y$y%y'y(y)y*y+y,y-y.y/y:y;y=y?y@y[y]y^y_y`y{y|y}y~yazbzczdzezfzgzhzizjzkzlzmznzozpzqzrzsztzuzvzwzxzyzzzAzBzCzDzEzFzGzHzIzJzKzLzMzNzOzPzQzRzSzTzUzVzWzXzYzZz0z1z2z3z4z5z6z7z8z9z!z#z$z%z'z(z)z*z+z,z-z.z/z:z;z=z?z@z[z]z^z_z`z{z|z}z~zaAbAcAdAeAfAgAhAiAjAkAlAmAnAoApAqArAsAtAuAvAwAxAyAzAAABACADAEAFAGAHAIAJAKALAMANAOAPAQARASATAUAVAWAXAYAZA0A1A2A3A4A5A6A7A8A9A!A#A$A%A'A(A)A*A+A,A-A.A/A:A;A=A?A@A[A]A^A_A`A{A|A}A~AaBbBcBdBeBfBgBhBiBjBkBlBmBnBoBpBqBrBsBtBuBvBwBxByBzBABBBCBDBEBFBGBHBIBJBKBLBMBNBOBPBQBRBSBTBUBVBWBXBYBZB0B1B2B3B4B5B6B7B8B9B!B#B$B%B'B(B)B*B+B,B-B.B/B:B;B=B?B@B[B]B^B_B`B{B|B}B~BaCbCcC0cdCeCfCgChCiCjCkClCmCnCoCpCqCrCsCtCuCvCwCxCyCzCACBCCCDCECFCGCHCICJCKCLCMCNCOCPCQCRCSCTCUCVCWCXCYCZC0CUh1C2C3C4C5C6C7C8C9C!C#C$C%C'C(C)C*C+C,C-C.C/C:C;C=C?C@C[C]C^C_C`C{C|C}C~CaDbDcDdDeDfDgDhDiDjDkDlDmDnDoDpDqDrDsDtDuDvDwDxDxfyDzDADBDCDDDEDFDGDHDIDJDKDLDMDNDODPDQDRDSDTDUDVDWDXDYDZD0D1D2D

166 cdef const char* name 

167 name_err = cydriver.cuGetErrorName(error, &name) 2chghhhihkhNelhyGmi{h|h}h~haisb1b5d

168 if name_err != cydriver.CUresult.CUDA_SUCCESS: 2chghhhihkhNelhyGmi{h|h}h~haisb1b5d

169 raise CUDAError(f"UNEXPECTED ERROR CODE: {error}") 2mi

170 with gil: 2chghhhihkhNelhyGmi{h|h}h~haisb1b5d

171 # TODO: consider lower this to Cython 

172 expl = DRIVER_CU_RESULT_EXPLANATIONS.get(int(error)) 2chghhhihkhNelhyGmi{h|h}h~haisb1b5d

173 if expl is not None: 2chghhhihkhNelhyGmi{h|h}h~haisb1b5d

174 raise CUDAError(f"{name.decode()}: {expl}") 2chghhhihkhNelhyGmi{h|h}h~haisb1b5d

175 cdef const char* desc 

176 desc_err = cydriver.cuGetErrorString(error, &desc) 

177 if desc_err != cydriver.CUresult.CUDA_SUCCESS: 

178 raise CUDAError(f"{name.decode()}") 

179 raise CUDAError(f"{name.decode()}: {desc.decode()}") 

180  

181  

182cpdef inline int _check_runtime_error(error) except?-1: 

183 if error == _RUNTIME_SUCCESS: 2nioipiqirisitiuiviwixiyiziAiBiCiDiEiFiGiHiIiJiKiLiMiNiOiPiQiRiSiTiUiViWiXiYiZi0i1i2i3i4i5i6i7i8i9i!i#i$iTbUb%i'i(i)i*i+iVbWbXbYbZb0bNguNVhOgTeUePgQgRg/b:btcucfbgbhb2bl m n o p q r s t u v w x y z A ibjbkb3bB C D E F G H I J K L M N O P Q lbmbnb4bR S T U V W X Y Z 0 1 2 3 4 5 6 [cIcWe]cXeYe^c0e1e4e5e

184 return 0 2nioipiqirisitiuiviwixiyiziAiBiCiDiEiFiGiHiIiJiKiLiMiNiOiPiQiRiSiTiUiViWiXiYiZi0i1i2i3i4i5i6i7i8i9i!i#i$iTbUb%i'i(i)i*i+iVbWbXbYbZb0bNgVhOgTeUePgQgRg/b:btcucfbgbhb2bl m n o p q r s t u v w x y z A ibjbkb3bB C D E F G H I J K L M N O P Q lbmbnb4bR S T U V W X Y Z 0 1 2 3 4 5 6 [cIcWe]cXeYe^c0e1e4e5e

185 # `_check_error()` reaches this path only for `runtime.cudaError_t` values. 

186 # Use the enum name directly because Windows hybrid cudart can lag that table. 

187 name = error.name 2uNVh

188 expl = RUNTIME_CUDA_ERROR_EXPLANATIONS.get(int(error)) 2uNVh

189 if expl is not None: 2uNVh

190 raise CUDAError(f"{name}: {expl}") 2uNVh

191 desc_err, desc = runtime.cudaGetErrorString(error) 2Vh

192 if desc_err != _RUNTIME_SUCCESS: 2Vh

193 raise CUDAError(f"{name}") 

194 desc = desc.decode() 2Vh

195 raise CUDAError(f"{name}: {desc}") 2Vh

196  

197  

198cpdef inline int _check_nvrtc_error(error, handle=None) except?-1: 

199 if error == _NVRTC_SUCCESS: 2a 3DvNb ;b=b?b@b[b]b^b_b`b{b|b}b~bacbcccdcecfcgchcicjckclcmcncocpcqcrcscvcNcOcPcWhQcXhRcYhScIgJgUcVcWcXc6d7d8d`c{c|c9d!d#d$d1c2c}c%d~c'd(d)d*dad+dbdcddded,d-d.d/d:d;d=d?d@dfdgdhdc YcwcZc[d4g5g6g7g8g9gZh0h1h2h3h4h5h6h7h8h9h!h#h$h%h'h(h)h*h+h,h-h.h/h:h;h=h]d^d_d`d{d|d}d~daebe

200 return 0 2a vNb ;b=b?b@b[b]b^b_b`b{b|b}b~bacbcccdcecfcgchcicjckclcmcncocpcqcrcscvcNcOcPcWhQcXhRcYhScIgJgUcVcWcXc6d7d8d`c{c|c9d!d#d$d1c2c}c%d~c'd(d)d*dad+dbdcddded,d-d.d/d:d;d=d?d@dfdgdhdc YcwcZc[d4g5g6g7g8g9gZh0h1h2h3h4h5h6h7h8h9h!h#h$h%h'h(h)h*h+h,h-h.h/h:h;h=h]d^d_d`d{d|d}d~daebe

201 err = f"{error}: {nvrtc.nvrtcGetErrorString(error)[1].decode()}" 23DvN

202 if handle is not None: 23DvN

203 _, logsize = nvrtc.nvrtcGetProgramLogSize(handle) 23D

204 log = b" " * logsize 23D

205 _ = nvrtc.nvrtcGetProgramLog(handle, log) 23D

206 err += f", compilation log:\n\n{log.decode('utf-8', errors='backslashreplace')}" 23D

207 raise NVRTCError(err) 23DvN

208  

209  

210cdef inline int _check_error(error, handle=None) except?-1: 

211 if isinstance(error, driver.CUresult): 2a Nj@eOj[ePj]eQj8 RjtbSjTjUjubVjWjXjvbYjwbZj0jxb1jyb2jzb3j9 4j! 5j# 6j$ 7j% 8j' 9j( !j) #j* $j+ %j, 'j- (j. )j/ *j: +j; ,j= -j? .j@ /j[ :j] ;j^ =j_ ?j` @j{ [j| ]j^j_j`j{j|j}j~jakbkckdkekfkgkhkikjkkklk^emknkok_epk`eqk{erk|eskAbtk}euk~evkafwkxkykBbzkCbAkDbBkEbCkFbDk(cEk)cFk*cGk+cHk,cIk-cJk.cKk/cLkMkNkOkPkQkRkSkdhTkehUkfhVkWkXkYkZk0k1k2k3k4k5k6k7k8k9k!k#k$k%k'k(k)k*k+k,k-k.k/k:k;k=k?k@k[k]k^k_k`k{k|k}k~kalblcldlelflglhliljlklllmlnlolplqlrlsltlulvlwlxlylzlAlBlClDlElFlGlHlIlJlKlLlMlNlOlPlQlRlSlTlUlVlWlXlYlZl0l1l2l3l4l5l6l7l8l9l!l#l$l%l'l(l)l*l+l,l-l.l/l:l;l=l?l@l[l]l^l_l`l{l|l}l~lambmcmdmemfmgmhmimjmkmlmmmnmompmqmrmsmtmumvmwmxmymzmAmBmCmDmEmFmGm6bHmImJmbfKmLmMmNmOmPmQmRmSmTmUmVmWmXmYmZmcf0mdf1m2m3m4m5m:c6m0d7m8m9mf !mg #m$m%mGb'mob(mHb)m*mef+mjh,mIb-mJb.mKb/mLb:mMb;m=m?m@m[md ]m^m_m`mOePeQeffgfhfh i j k } ~ abbbpbqbrbifjfkflfmfnf{mcb|mNb}mOb~mPbanQbbnRbcndnSbendbfngnofpfhninmhnhciohphdiqhrheishthfiuhgijnLgknqfhilnRemnrfsfiitfnnufjionMgpnvhkiwhxhlivfqnwfrnsntnunvnwnSexnnioipiqirisitiuiviwixiyi7 zie AiHcBiCiDiEiFiGiHiIiJiKiLiMiNiOiPiQiRiSiTiUiViWiXiYiZiFg0i1i2i3i4i5iGg6i7i8i1d9i2d!i;c#i=c$iTbUb?c%i@c'i(i)i*i+iVbWbXbYbZb0bNgynznAnBnCnDnEnFnOgTeGnUeHnInJnKnLnMnNnOnPgPnQnRnSnChTnUnVnDhQgRgWnEhXnYnZn0n1n2n3n4n5n6n7n8n9n!n#n$n%n'nFhGhHh(n)nIhJhKh*nLhMhNh+n,n-n.n/n:n;n=n?n@n[n]n^n_n`n{n|n}n~neb/b:baobocodob ;beo=bfo?bgo@bho[bio]bjo^bko_blo`bmo{bno|boo}bpo~bqoacrobcsocctodcuoecvofcwogcxohcyoiczojcAokcBolcComcDoncEoocFopcGoqcHorcIoJoscKoLoMoNoSgtcucTgUgVgWgXgYgZg0gOhPhce3cfbxcgb4cychb2b6e7e8el %bm n o p q r s t u 'bv w x y 7bz A 8b(bde5cibzcjb6cAckb3b9e!e#eB )bC D E F G H I J K *bL M N O 9bP Q !b+bee7clbBcmb8cCcnb4b$e%e'eR ,bS T U V W X Y Z 0 -b1 2 3 4 #b5 6 $b.b1g[cOoPoQoVeRoSoToUoVoWoXoIcYoWeZo]c0oXe1o2o3o4o5o6o7o8o9o!o#o$o%o'o(o)o*o+o,o-o.o/o:o;o=o?o@o[o]o^o_o`o{o|o}o~oapbpcpdpepfpgphpipjpkplpmpnpopppqprpspYe2g3gZetp^cup0e1evpwpxpypzpApBpvcCpDpEp3dFpGpHpIpJpKpLpMp1bNpOpPpQpRpSpTpUpVpWpXpYpZp0pNc1pOc2pPc3p4p5p6pWhQc7pXhRc8pYhSc9p!p#p$p%p'p(p)p*p+p,p-p.p/p:p;p=p?p@p[p]p^p_p`p{p|p}p~paqbqcqdqeqfqgqhqiqjqkqlqmqnqoqpqqqrqsqtquqvqwqxqyqzqAqBqCqDqEqFqGqHqIqJqKqLqMqNqOqPqQqRqSqTqUqVqWqXqYqZq0q1q2q3q4q5q6q7q8q9q!q#q$q%q'q(q)q*q+q,q-q.q/q:q;q=q?q@q[q]q^q_q`q{q|q}q~qarbrcrdrerfrgrhrirjrkrlrmrnrorprqrrrsrtrurvrwrxryrzrArBrCrDrErFrGrHrIrJrKrLrMrNrOrPrQrRrSrTrUrVrWrXrYrZr0r1r2r3r4r5r6r7r8r9r!r#r$r%r'r(r)r*r+r,r-r.r/r:r;r=r?r@r[r]r^r_r`r{r|r}r~rasbscsdsesfsgshsisjskslsmsnsospsqsrssstsusvswsxsyszsAsBsCsDsEsFsGsHsIsJsKsLsMsNsOsPsQsRsSsTsUsVsWsXsYsZs0s1s2s3s4s5s6s7s8s9s!s#s$s%s's(s)s*s+s,s-s.s/s:s;s=s?s@s[s]s^s_s`s{s|s}s~satbtctdtetftgthtitjtktltmtntotptqtrtstttutvtwtxtytztAtBtCtDtEtFtGtHtItJtKtLtMtNtOtPtQtRtStTtUtVtWtXtYtZt0t1t2t3t4t5t6t7t8t9t!t#t$t%t't(t)t*t+t,t-t.t/t:t;t=t?t@t[t]t^t_t`t{t|t}t~taubucudueufuguhuiujukulumunuoupuqurusutuuuvuwuxuyuzuAuBuCuDuEuFuGuHuIuJuKuLuMuNuOuPuQuRuSuTuUuVuWuXuYuZu0u1u2u3u4u5u6u7u8u9u!u#u$u%u'u(u)u*u+u,u-u.u/u:u;u=u?u@u[u]u^u_u`u{u|u}u~uavbvcvdvevfvgvhvivjvkvlvmvnvovpvqvrvsvtvuvvvwvxvyvzvAvBvCvDvEvFvGvHvIvJvKvLvMvNvOvPvQvRvSvTvUvVvWvXvYvZv0v1v2v3v4v5v6v7v8v9v!v#v$v%v'v(v)v*v+v,v-v.v/v:v;v=v?v@v[v]v^v_v`v{v|v}v~vawbwcwdwewfwgwhwiwjwkwlwmwnwowpwqwrwswtwuwvwwwxwywzwAwBwCwDwEwFwGwHwIwJwKwLwMwNwOwPwQwRwSwTwUwVwWwXwYwZw0w1w2w3w4w5w6w7w8w9w!w#w$w%w'w(w)w*w+w,w-w.w/w:w;w=w?w@w[w]w^w_w`w{w|w}w~waxbxcxdxexfxgxhxixjxkxlxmxnxoxpxqxrxsxtxuxvxwxxxyxzxAxBxCxDxExFxGxHxIxJxKxLxMxNxOxPxQxRxSxTxUxVxWxXxYxZx0x1x2x3x4x5x6x7x8x9x!x#x$x%x'x(x)x*x+x,x-x.x/x:x;x=x?x@x[x]x^x_x`x{x|x}x~xaybycydyey4e5eTcfygyhyiyjykylymynyoypyqyrysytyuyvywyxyyyzyAyByIgCyDyEyJgFyGyHyIyJyKyLyMyNyOyPyQyRySyTyUyVyWyXyYyZy0y1y2y3y4y5y6y7y8y9y!y#y$y%y'y(y)y*y+y,y-y.y/y:y;y=y?y@y[y]y^y_y`y{y|y}y~yazbzczdzezfzgzhzizjzkzlzmznzozpzqzrzsztzuzvzwzxzyzzzAzBzCzDzEzFzGzHzIzJzKzLzMzNzOzPzQzRzSzTzUzVzWzXzYzZz0z1z2z3z4z5z6z7z8z9z!z#z$z%z'z(z)z*z+z,z-z.z/z:z;z=z?z@z[z]z^z_z`z{z|z}z~zaAbAcAdAeAfAgAhAiAjAkAlAmAnAoApAqArAsAtAuAvAwAxAyAzAAABAUcCAVcDAWcEAXcFA6dGA7dHA8dIA`cJA{cKA|cLA9dMA!dNA#dOA$dPA1cQA2cRA}cSA%dTA~cUA'dVA(dWA)dXA*dYAadZA+d0Abd1Acd2Add3Aed4A,d5A-d6A.d7A/d8A:d9A;d!A=d#A?d$A@d%Afd'Agd(Ahd)Ac *AYc+A,A-A.A/A:A;A=A?A@A[A]A^A_A`Awc{A|A}A~AaBZcbBcBdBeBfBgBhB[diB4g5g6g7g8g9gZh0h1h2h3h4h5h6h7h8h9h!h#h$hjB%h'h(h)h*h+h,h-h.h/hkBlBmBnB:h;h=h]d^d_d`d{doB|d}d~daebepBqBrBsBtBuBvBwBxByBzBABBBCBDBEBFBGBHBIBJBKBLBMBNBOBPBQBRBSBTBUBVBWBXBYBZB0B1B2B3B4B5B6B7B8B9B!B#B$B%B'B(B)B*B+B,B-B.B/B:B;B=B?B@B[B]B^B_B`B{B|B}B~BaCbCcC0cdCeCfCgChCiCjCkClCmCnCoCpCqCrCsCtCuCvCwCxCyCzCACBCCCDCECFCGCHCICJCKCLCMCNCOCPCQCRCSCTCUCVCWCXCYCZC0CUh1C2C3C4C5C6C7C8C9C!C#C$C%C'C(C)C*C+C,C-C.C/C:C;C=C?C@C[C]C^C_C`C{C|C}C~CaDbDcDdDeDfDgDhDiDjDkDlDmDnDoDpDqDrDsDtDuDvDwDxDxfyDzDADBDCDDDEDFDGDHDIDJDKDLDMDNDODPDQDRDSDTDUDVDWDXDYDZD0D1D2D

212 return _check_driver_error(error) 2a Nj@eOj[ePj]eQj8 RjtbSjTjUjubVjWjXjvbYjwbZj0jxb1jyb2jzb3j9 4j! 5j# 6j$ 7j% 8j' 9j( !j) #j* $j+ %j, 'j- (j. )j/ *j: +j; ,j= -j? .j@ /j[ :j] ;j^ =j_ ?j` @j{ [j| ]j^j_j`j{j|j}j~jakbkckdkekfkgkhkikjkkklk^emknkok_epk`eqk{erk|eskAbtk}euk~evkafwkxkykBbzkCbAkDbBkEbCkFbDk(cEk)cFk*cGk+cHk,cIk-cJk.cKk/cLkMkNkOkPkQkRkSkdhTkehUkfhVkWkXkYkZk0k1k2k3k4k5k6k7k8k9k!k#k$k%k'k(k)k*k+k,k-k.k/k:k;k=k?k@k[k]k^k_k`k{k|k}k~kalblcldlelflglhliljlklllmlnlolplqlrlsltlulvlwlxlylzlAlBlClDlElFlGlHlIlJlKlLlMlNlOlPlQlRlSlTlUlVlWlXlYlZl0l1l2l3l4l5l6l7l8l9l!l#l$l%l'l(l)l*l+l,l-l.l/l:l;l=l?l@l[l]l^l_l`l{l|l}l~lambmcmdmemfmgmhmimjmkmlmmmnmompmqmrmsmtmumvmwmxmymzmAmBmCmDmEmFmGm6bHmImJmbfKmLmMmNmOmPmQmRmSmTmUmVmWmXmYmZmcf0mdf1m2m3m4m5m:c6m0d7m8m9mf !mg #m$m%mGb'mob(mHb)m*mef+mjh,mIb-mJb.mKb/mLb:mMb;m=m?m@m[md ]m^m_m`mOePeQeffgfhfh i j k } ~ abbbpbqbrbifjfkflfmfnf{mcb|mNb}mOb~mPbanQbbnRbcndnSbendbfngnofpfhninmhnhciohphdiqhrheishthfiuhgijnLgknqfhilnRemnrfsfiitfnnufjionMgpnvhkiwhxhlivfqnwfrnsntnunvnwnSexn7 e HcFgGg1d2d;c=cTbUb?c@cVbWbXbYbZb0bynznAnBnCnDnEnFnGnHnInJnKnLnMnNnOnPnQnRnSnChTnUnVnDhWnEhXnYnZn0n1n2n3n4n5n6n7n8n9n!n#n$n%n'nFhGhHh(n)nIhJhKh*nLhMhNh+n,n-n.n/n:n;n=n?n@n[n]n^n_n`n{n|n}n~neb/b:baobocodoeofogohoiojokolomonooopoqorosotouovowoxoyozoAoBoCoDoEoFoGoHoIoJoKoLoMoNoSgtcucTgUgVgWgXgYgZg0gOhPhce3cfbxcgb4cychb2b6e7e8el %bm n o p q r s t u 'bv w x y 7bz A 8b(bde5cibzcjb6cAckb3b9e!e#eB )bC D E F G H I J K *bL M N O 9bP Q !b+bee7clbBcmb8cCcnb4b$e%e'eR ,bS T U V W X Y Z 0 -b1 2 3 4 #b5 6 $b.b1gOoPoQoVeRoSoToUoVoWoXoYoZo0o1o2o3o4o5o6o7o8o9o!o#o$o%o'o(o)o*o+o,o-o.o/o:o;o=o?o@o[o]o^o_o`o{o|o}o~oapbpcpdpepfpgphpipjpkplpmpnpopppqprpsp2g3gZetpupvpwpxpypzpApBpCpDpEp3dFpGpHpIpJpKpLpMp1bNpOpPpQpRpSpTpUpVpWpXpYpZp0p1p2p3p4p5p6p7p8p9p!p#p$p%p'p(p)p*p+p,p-p.p/p:p;p=p?p@p[p]p^p_p`p{p|p}p~paqbqcqdqeqfqgqhqiqjqkqlqmqnqoqpqqqrqsqtquqvqwqxqyqzqAqBqCqDqEqFqGqHqIqJqKqLqMqNqOqPqQqRqSqTqUqVqWqXqYqZq0q1q2q3q4q5q6q7q8q9q!q#q$q%q'q(q)q*q+q,q-q.q/q:q;q=q?q@q[q]q^q_q`q{q|q}q~qarbrcrdrerfrgrhrirjrkrlrmrnrorprqrrrsrtrurvrwrxryrzrArBrCrDrErFrGrHrIrJrKrLrMrNrOrPrQrRrSrTrUrVrWrXrYrZr0r1r2r3r4r5r6r7r8r9r!r#r$r%r'r(r)r*r+r,r-r.r/r:r;r=r?r@r[r]r^r_r`r{r|r}r~rasbscsdsesfsgshsisjskslsmsnsospsqsrssstsusvswsxsyszsAsBsCsDsEsFsGsHsIsJsKsLsMsNsOsPsQsRsSsTsUsVsWsXsYsZs0s1s2s3s4s5s6s7s8s9s!s#s$s%s's(s)s*s+s,s-s.s/s:s;s=s?s@s[s]s^s_s`s{s|s}s~satbtctdtetftgthtitjtktltmtntotptqtrtstttutvtwtxtytztAtBtCtDtEtFtGtHtItJtKtLtMtNtOtPtQtRtStTtUtVtWtXtYtZt0t1t2t3t4t5t6t7t8t9t!t#t$t%t't(t)t*t+t,t-t.t/t:t;t=t?t@t[t]t^t_t`t{t|t}t~taubucudueufuguhuiujukulumunuoupuqurusutuuuvuwuxuyuzuAuBuCuDuEuFuGuHuIuJuKuLuMuNuOuPuQuRuSuTuUuVuWuXuYuZu0u1u2u3u4u5u6u7u8u9u!u#u$u%u'u(u)u*u+u,u-u.u/u:u;u=u?u@u[u]u^u_u`u{u|u}u~uavbvcvdvevfvgvhvivjvkvlvmvnvovpvqvrvsvtvuvvvwvxvyvzvAvBvCvDvEvFvGvHvIvJvKvLvMvNvOvPvQvRvSvTvUvVvWvXvYvZv0v1v2v3v4v5v6v7v8v9v!v#v$v%v'v(v)v*v+v,v-v.v/v:v;v=v?v@v[v]v^v_v`v{v|v}v~vawbwcwdwewfwgwhwiwjwkwlwmwnwowpwqwrwswtwuwvwwwxwywzwAwBwCwDwEwFwGwHwIwJwKwLwMwNwOwPwQwRwSwTwUwVwWwXwYwZw0w1w2w3w4w5w6w7w8w9w!w#w$w%w'w(w)w*w+w,w-w.w/w:w;w=w?w@w[w]w^w_w`w{w|w}w~waxbxcxdxexfxgxhxixjxkxlxmxnxoxpxqxrxsxtxuxvxwxxxyxzxAxBxCxDxExFxGxHxIxJxKxLxMxNxOxPxQxRxSxTxUxVxWxXxYxZx0x1x2x3x4x5x6x7x8x9x!x#x$x%x'x(x)x*x+x,x-x.x/x:x;x=x?x@x[x]x^x_x`x{x|x}x~xaybycydyeyTcfygyhyiyjykylymynyoypyqyrysytyuyvywyxyyyzyAyByCyDyEyFyGyHyIyJyKyLyMyNyOyPyQyRySyTyUyVyWyXyYyZy0y1y2y3y4y5y6y7y8y9y!y#y$y%y'y(y)y*y+y,y-y.y/y:y;y=y?y@y[y]y^y_y`y{y|y}y~yazbzczdzezfzgzhzizjzkzlzmznzozpzqzrzsztzuzvzwzxzyzzzAzBzCzDzEzFzGzHzIzJzKzLzMzNzOzPzQzRzSzTzUzVzWzXzYzZz0z1z2z3z4z5z6z7z8z9z!z#z$z%z'z(z)z*z+z,z-z.z/z:z;z=z?z@z[z]z^z_z`z{z|z}z~zaAbAcAdAeAfAgAhAiAjAkAlAmAnAoApAqArAsAtAuAvAwAxAyAzAAABACADAEAFAGAHAIAJAKALAMANAOAPAQARASATAUAVAWAXAYAZA0A1A2A3A4A5A6A7A8A9A!A#A$A%A'A(A)A*A+A,A-A.A/A:A;A=A?A@A[A]A^A_A`A{A|A}A~AaBbBcBdBeBfBgBhBiBjBkBlBmBnBoBpBqBrBsBtBuBvBwBxByBzBABBBCBDBEBFBGBHBIBJBKBLBMBNBOBPBQBRBSBTBUBVBWBXBYBZB0B1B2B3B4B5B6B7B8B9B!B#B$B%B'B(B)B*B+B,B-B.B/B:B;B=B?B@B[B]B^B_B`B{B|B}B~BaCbCcC0cdCeCfCgChCiCjCkClCmCnCoCpCqCrCsCtCuCvCwCxCyCzCACBCCCDCECFCGCHCICJCKCLCMCNCOCPCQCRCSCTCUCVCWCXCYCZC0CUh1C2C3C4C5C6C7C8C9C!C#C$C%C'C(C)C*C+C,C-C.C/C:C;C=C?C@C[C]C^C_C`C{C|C}C~CaDbDcDdDeDfDgDhDiDjDkDlDmDnDoDpDqDrDsDtDuDvDwDxDxfyDzDADBDCDDDEDFDGDHDIDJDKDLDMDNDODPDQDRDSDTDUDVDWDXDYDZD0D1D2D

213 elif isinstance(error, runtime.cudaError_t): 2a nioipiqirisitiuiviwixiyiziAiBiCiDiEiFiGiHiIiJiKiLiMiNiOiPiQiRiSiTiUiViWiXiYiZi0i1i2i3i4i5i6i7i8i9i!i#i$iTbUb%i'i(i)i*i+iVbWbXbYbZb0bNgOgTeUePgQgRg/b:bb ;b=b?b@b[b]b^b_b`b{b|b}b~bacbcccdcecfcgchcicjckclcmcncocpcqcrcsctcucfbgbhb2bl m n o p q r s t u v w x y z A ibjbkb3bB C D E F G H I J K L M N O P Q lbmbnb4bR S T U V W X Y Z 0 1 2 3 4 5 6 [cIcWe]cXeYe^c0e1evcNcOcPcWhQcXhRcYhSc4e5eIgJgUcVcWcXc6d7d8d`c{c|c9d!d#d$d1c2c}c%d~c'd(d)d*dad+dbdcddded,d-d.d/d:d;d=d?d@dfdgdhdc YcwcZc[d4g5g6g7g8g9gZh0h1h2h3h4h5h6h7h8h9h!h#h$h%h'h(h)h*h+h,h-h.h/h:h;h=h]d^d_d`d{d|d}d~daebe

214 return _check_runtime_error(error) 2nioipiqirisitiuiviwixiyiziAiBiCiDiEiFiGiHiIiJiKiLiMiNiOiPiQiRiSiTiUiViWiXiYiZi0i1i2i3i4i5i6i7i8i9i!i#i$iTbUb%i'i(i)i*i+iVbWbXbYbZb0bNgOgTeUePgQgRg/b:btcucfbgbhb2bl m n o p q r s t u v w x y z A ibjbkb3bB C D E F G H I J K L M N O P Q lbmbnb4bR S T U V W X Y Z 0 1 2 3 4 5 6 [cIcWe]cXeYe^c0e1e4e5e

215 elif isinstance(error, nvrtc.nvrtcResult): 2a b ;b=b?b@b[b]b^b_b`b{b|b}b~bacbcccdcecfcgchcicjckclcmcncocpcqcrcscvcNcOcPcWhQcXhRcYhScIgJgUcVcWcXc6d7d8d`c{c|c9d!d#d$d1c2c}c%d~c'd(d)d*dad+dbdcddded,d-d.d/d:d;d=d?d@dfdgdhdc YcwcZc[d4g5g6g7g8g9gZh0h1h2h3h4h5h6h7h8h9h!h#h$h%h'h(h)h*h+h,h-h.h/h:h;h=h]d^d_d`d{d|d}d~daebe

216 return _check_nvrtc_error(error, handle=handle) 2a b ;b=b?b@b[b]b^b_b`b{b|b}b~bacbcccdcecfcgchcicjckclcmcncocpcqcrcscvcNcOcPcWhQcXhRcYhScIgJgUcVcWcXc6d7d8d`c{c|c9d!d#d$d1c2c}c%d~c'd(d)d*dad+dbdcddded,d-d.d/d:d;d=d?d@dfdgdhdc YcwcZc[d4g5g6g7g8g9gZh0h1h2h3h4h5h6h7h8h9h!h#h$h%h'h(h)h*h+h,h-h.h/h:h;h=h]d^d_d`d{d|d}d~daebe

217 else: 

218 raise RuntimeError(f"Unknown error type: {error}") 

219  

220  

221def handle_return(result: tuple[Any, ...], handle: object = None) -> Any: 

222 _check_error(result[0], handle=handle) 2a Nj@eOj[ePj]eQj8 RjtbSjTjUjubVjWjXjvbYjwbZj0jxb1jyb2jzb3j9 4j! 5j# 6j$ 7j% 8j' 9j( !j) #j* $j+ %j, 'j- (j. )j/ *j: +j; ,j= -j? .j@ /j[ :j] ;j^ =j_ ?j` @j{ [j| ]j^j_j`j{j|j}j~jakbkckdkekfkgkhkikjkkklk^emknkok_epk`eqk{erk|eskAbtk}euk~evkafwkxkykBbzkCbAkDbBkEbCkFbDk(cEk)cFk*cGk+cHk,cIk-cJk.cKk/cLkMkNkOkPkQkRkSkdhTkehUkfhVkWkXkYkZk0k1k2k3k4k5k6k7k8k9k!k#k$k%k'k(k)k*k+k,k-k.k/k:k;k=k?k@k[k]k^k_k`k{k|k}k~kalblcldlelflglhliljlklllmlnlolplqlrlsltlulvlwlxlylzlAlBlClDlElFlGlHlIlJlKlLlMlNlOlPlQlRlSlTlUlVlWlXlYlZl0l1l2l3l4l5l6l7l8l9l!l#l$l%l'l(l)l*l+l,l-l.l/l:l;l=l?l@l[l]l^l_l`l{l|l}l~lambmcmdmemfmgmhmimjmkmlmmmnmompmqmrmsmtmumvmwmxmymzmAmBmCmDmEmFmGm6bHmImJmbfKmLmMmNmOmPmQmRmSmTmUmVmWmXmYmZmcf0mdf1m2m3m4m5m:c6m0d7m8m9mf !mg #m$m%mGb'mob(mHb)m*mef+mjh,mIb-mJb.mKb/mLb:mMb;m=m?m@m[md ]m^m_m`mOePeQeffgfhfh i j k } ~ abbbpbqbrbifjfkflfmfnf{mcb|mNb}mOb~mPbanQbbnRbcndnSbendbfngnofpfhninmhnhciohphdiqhrheishthfiuhgijnLgknqfhilnRemnrfsfiitfnnufjionMgpnvhkiwhxhlivfqnwfrnsntnunvnwnSexnnioipiqirisitiuiviwixiyi7 zie AiHcBiCiDiEiFiGiHiIiJiKiLiMiNiOiPiQiRiSiTiUiViWiXiYiZiFg0i1i2i3i4i5iGg6i7i8i1d9i2d!i;c#i=c$iTbUb?c%i@c'i(i)i*i+iVbWbXbYbZb0bNgynznAnBnCnDnEnFnOgTeGnUeHnInJnKnLnMnNnOnPgPnQnRnSnChTnUnVnDhQgRgWnEhXnYnZn0n1n2n3n4n5n6n7n8n9n!n#n$n%n'nFhGhHh(n)nIhJhKh*nLhMhNh+n,n-n.n/n:n;n=n?n@n[n]n^n_n`n{n|n}n~neb/b:baobocodob ;beo=bfo?bgo@bho[bio]bjo^bko_blo`bmo{bno|boo}bpo~bqoacrobcsocctodcuoecvofcwogcxohcyoiczojcAokcBolcComcDoncEoocFopcGoqcHorcIoJoscKoLoMoNoSgtcucTgUgVgWgXgYgZg0gOhPhce3cfbxcgb4cychb2b6e7e8el %bm n o p q r s t u 'bv w x y 7bz A 8b(bde5cibzcjb6cAckb3b9e!e#eB )bC D E F G H I J K *bL M N O 9bP Q !b+bee7clbBcmb8cCcnb4b$e%e'eR ,bS T U V W X Y Z 0 -b1 2 3 4 #b5 6 $b.b1g[cOoPoQoVeRoSoToUoVoWoXoIcYoWeZo]c0oXe1o2o3o4o5o6o7o8o9o!o#o$o%o'o(o)o*o+o,o-o.o/o:o;o=o?o@o[o]o^o_o`o{o|o}o~oapbpcpdpepfpgphpipjpkplpmpnpopppqprpspYe2g3gZetp^cup0e1evpwpxpypzpApBpvcCpDpEp3dFpGpHpIpJpKpLpMp1bNpOpPpQpRpSpTpUpVpWpXpYpZp0pNc1pOc2pPc3p4p5p6pWhQc7pXhRc8pYhSc9p!p#p$p%p'p(p)p*p+p,p-p.p/p:p;p=p?p@p[p]p^p_p`p{p|p}p~paqbqcqdqeqfqgqhqiqjqkqlqmqnqoqpqqqrqsqtquqvqwqxqyqzqAqBqCqDqEqFqGqHqIqJqKqLqMqNqOqPqQqRqSqTqUqVqWqXqYqZq0q1q2q3q4q5q6q7q8q9q!q#q$q%q'q(q)q*q+q,q-q.q/q:q;q=q?q@q[q]q^q_q`q{q|q}q~qarbrcrdrerfrgrhrirjrkrlrmrnrorprqrrrsrtrurvrwrxryrzrArBrCrDrErFrGrHrIrJrKrLrMrNrOrPrQrRrSrTrUrVrWrXrYrZr0r1r2r3r4r5r6r7r8r9r!r#r$r%r'r(r)r*r+r,r-r.r/r:r;r=r?r@r[r]r^r_r`r{r|r}r~rasbscsdsesfsgshsisjskslsmsnsospsqsrssstsusvswsxsyszsAsBsCsDsEsFsGsHsIsJsKsLsMsNsOsPsQsRsSsTsUsVsWsXsYsZs0s1s2s3s4s5s6s7s8s9s!s#s$s%s's(s)s*s+s,s-s.s/s:s;s=s?s@s[s]s^s_s`s{s|s}s~satbtctdtetftgthtitjtktltmtntotptqtrtstttutvtwtxtytztAtBtCtDtEtFtGtHtItJtKtLtMtNtOtPtQtRtStTtUtVtWtXtYtZt0t1t2t3t4t5t6t7t8t9t!t#t$t%t't(t)t*t+t,t-t.t/t:t;t=t?t@t[t]t^t_t`t{t|t}t~taubucudueufuguhuiujukulumunuoupuqurusutuuuvuwuxuyuzuAuBuCuDuEuFuGuHuIuJuKuLuMuNuOuPuQuRuSuTuUuVuWuXuYuZu0u1u2u3u4u5u6u7u8u9u!u#u$u%u'u(u)u*u+u,u-u.u/u:u;u=u?u@u[u]u^u_u`u{u|u}u~uavbvcvdvevfvgvhvivjvkvlvmvnvovpvqvrvsvtvuvvvwvxvyvzvAvBvCvDvEvFvGvHvIvJvKvLvMvNvOvPvQvRvSvTvUvVvWvXvYvZv0v1v2v3v4v5v6v7v8v9v!v#v$v%v'v(v)v*v+v,v-v.v/v:v;v=v?v@v[v]v^v_v`v{v|v}v~vawbwcwdwewfwgwhwiwjwkwlwmwnwowpwqwrwswtwuwvwwwxwywzwAwBwCwDwEwFwGwHwIwJwKwLwMwNwOwPwQwRwSwTwUwVwWwXwYwZw0w1w2w3w4w5w6w7w8w9w!w#w$w%w'w(w)w*w+w,w-w.w/w:w;w=w?w@w[w]w^w_w`w{w|w}w~waxbxcxdxexfxgxhxixjxkxlxmxnxoxpxqxrxsxtxuxvxwxxxyxzxAxBxCxDxExFxGxHxIxJxKxLxMxNxOxPxQxRxSxTxUxVxWxXxYxZx0x1x2x3x4x5x6x7x8x9x!x#x$x%x'x(x)x*x+x,x-x.x/x:x;x=x?x@x[x]x^x_x`x{x|x}x~xaybycydyey4e5eTcfygyhyiyjykylymynyoypyqyrysytyuyvywyxyyyzyAyByIgCyDyEyJgFyGyHyIyJyKyLyMyNyOyPyQyRySyTyUyVyWyXyYyZy0y1y2y3y4y5y6y7y8y9y!y#y$y%y'y(y)y*y+y,y-y.y/y:y;y=y?y@y[y]y^y_y`y{y|y}y~yazbzczdzezfzgzhzizjzkzlzmznzozpzqzrzsztzuzvzwzxzyzzzAzBzCzDzEzFzGzHzIzJzKzLzMzNzOzPzQzRzSzTzUzVzWzXzYzZz0z1z2z3z4z5z6z7z8z9z!z#z$z%z'z(z)z*z+z,z-z.z/z:z;z=z?z@z[z]z^z_z`z{z|z}z~zaAbAcAdAeAfAgAhAiAjAkAlAmAnAoApAqArAsAtAuAvAwAxAyAzAAABAUcCAVcDAWcEAXcFA6dGA7dHA8dIA`cJA{cKA|cLA9dMA!dNA#dOA$dPA1cQA2cRA}cSA%dTA~cUA'dVA(dWA)dXA*dYAadZA+d0Abd1Acd2Add3Aed4A,d5A-d6A.d7A/d8A:d9A;d!A=d#A?d$A@d%Afd'Agd(Ahd)Ac *AYc+A,A-A.A/A:A;A=A?A@A[A]A^A_A`Awc{A|A}A~AaBZcbBcBdBeBfBgBhB[diB4g5g6g7g8g9gZh0h1h2h3h4h5h6h7h8h9h!h#h$hjB%h'h(h)h*h+h,h-h.h/hkBlBmBnB:h;h=h]d^d_d`d{doB|d}d~daebepBqBrBsBtBuBvBwBxByBzBABBBCBDBEBFBGBHBIBJBKBLBMBNBOBPBQBRBSBTBUBVBWBXBYBZB0B1B2B3B4B5B6B7B8B9B!B#B$B%B'B(B)B*B+B,B-B.B/B:B;B=B?B@B[B]B^B_B`B{B|B}B~BaCbCcC0cdCeCfCgChCiCjCkClCmCnCoCpCqCrCsCtCuCvCwCxCyCzCACBCCCDCECFCGCHCICJCKCLCMCNCOCPCQCRCSCTCUCVCWCXCYCZC0CUh1C2C3C4C5C6C7C8C9C!C#C$C%C'C(C)C*C+C,C-C.C/C:C;C=C?C@C[C]C^C_C`C{C|C}C~CaDbDcDdDeDfDgDhDiDjDkDlDmDnDoDpDqDrDsDtDuDvDwDxDxfyDzDADBDCDDDEDFDGDHDIDJDKDLDMDNDODPDQDRDSDTDUDVDWDXDYDZD0D1D2D

223 cdef int out_len = len(result) 2a Nj@eOj[ePj]eQj8 RjtbSjTjUjubVjWjXjvbYjwbZj0jxb1jyb2jzb3j9 4j! 5j# 6j$ 7j% 8j' 9j( !j) #j* $j+ %j, 'j- (j. )j/ *j: +j; ,j= -j? .j@ /j[ :j] ;j^ =j_ ?j` @j{ [j| ]j^j_j`j{j|j}j~jakbkckdkekfkgkhkikjkkklk^emknkok_epk`eqk{erk|eskAbtk}euk~evkafwkxkykBbzkCbAkDbBkEbCkFbDk(cEk)cFk*cGk+cHk,cIk-cJk.cKk/cLkMkNkOkPkQkRkSkdhTkehUkfhVkWkXkYkZk0k1k2k3k4k5k6k7k8k9k!k#k$k%k'k(k)k*k+k,k-k.k/k:k;k=k?k@k[k]k^k_k`k{k|k}k~kalblcldlelflglhliljlklllmlnlolplqlrlsltlulvlwlxlylzlAlBlClDlElFlGlHlIlJlKlLlMlNlOlPlQlRlSlTlUlVlWlXlYlZl0l1l2l3l4l5l6l7l8l9l!l#l$l%l'l(l)l*l+l,l-l.l/l:l;l=l?l@l[l]l^l_l`l{l|l}l~lambmcmdmemfmgmhmimjmkmlmmmnmompmqmrmsmtmumvmwmxmymzmAmBmCmDmEmFmGm6bHmImJmbfKmLmMmNmOmPmQmRmSmTmUmVmWmXmYmZmcf0mdf1m2m3m4m5m:c6m0d7m8m9mf !mg #m$m%mGb'mob(mHb)m*mef+mjh,mIb-mJb.mKb/mLb:mMb;m=m?m@m[md ]m^m_m`mOePeQeffgfhfh i j k } ~ abbbpbqbrbifjfkflfmfnf{mcb|mNb}mOb~mPbanQbbnRbcndnSbendbfngnofpfhninmhnhciohphdiqhrheishthfiuhgijnLgknqfhilnRemnrfsfiitfnnufjionMgpnvhkiwhxhlivfqnwfrnsntnunvnwnSexnnioipiqirisitiuiviwixiyi7 zie AiHcBiCiDiEiFiGiHiIiJiKiLiMiNiOiPiQiRiSiTiUiViWiXiYiZiFg0i1i2i3i4i5iGg6i7i8i1d9i2d!i;c#i=c$iTbUb?c%i@c'i(i)i*i+iVbWbXbYbZb0bNgynznAnBnCnDnEnFnOgTeGnUeHnInJnKnLnMnNnOnPgPnQnRnSnChTnUnVnDhQgRgWnEhXnYnZn0n1n2n3n4n5n6n7n8n9n!n#n$n%n'nFhGhHh(n)nIhJhKh*nLhMhNh+n,n-n.n/n:n;n=n?n@n[n]n^n_n`n{n|n}n~neb/b:baobocodob ;beo=bfo?bgo@bho[bio]bjo^bko_blo`bmo{bno|boo}bpo~bqoacrobcsocctodcuoecvofcwogcxohcyoiczojcAokcBolcComcDoncEoocFopcGoqcHorcIoJoscKoLoMoNoSgtcucTgUgVgWgXgYgZg0gOhPhce3cfbxcgb4cychb2b6e7e8el %bm n o p q r s t u 'bv w x y 7bz A 8b(bde5cibzcjb6cAckb3b9e!e#eB )bC D E F G H I J K *bL M N O 9bP Q !b+bee7clbBcmb8cCcnb4b$e%e'eR ,bS T U V W X Y Z 0 -b1 2 3 4 #b5 6 $b.b1g[cOoPoQoVeRoSoToUoVoWoXoIcYoWeZo]c0oXe1o2o3o4o5o6o7o8o9o!o#o$o%o'o(o)o*o+o,o-o.o/o:o;o=o?o@o[o]o^o_o`o{o|o}o~oapbpcpdpepfpgphpipjpkplpmpnpopppqprpspYe2g3gZetp^cup0e1evpwpxpypzpApBpvcCpDpEp3dFpGpHpIpJpKpLpMp1bNpOpPpQpRpSpTpUpVpWpXpYpZp0pNc1pOc2pPc3p4p5p6pWhQc7pXhRc8pYhSc9p!p#p$p%p'p(p)p*p+p,p-p.p/p:p;p=p?p@p[p]p^p_p`p{p|p}p~paqbqcqdqeqfqgqhqiqjqkqlqmqnqoqpqqqrqsqtquqvqwqxqyqzqAqBqCqDqEqFqGqHqIqJqKqLqMqNqOqPqQqRqSqTqUqVqWqXqYqZq0q1q2q3q4q5q6q7q8q9q!q#q$q%q'q(q)q*q+q,q-q.q/q:q;q=q?q@q[q]q^q_q`q{q|q}q~qarbrcrdrerfrgrhrirjrkrlrmrnrorprqrrrsrtrurvrwrxryrzrArBrCrDrErFrGrHrIrJrKrLrMrNrOrPrQrRrSrTrUrVrWrXrYrZr0r1r2r3r4r5r6r7r8r9r!r#r$r%r'r(r)r*r+r,r-r.r/r:r;r=r?r@r[r]r^r_r`r{r|r}r~rasbscsdsesfsgshsisjskslsmsnsospsqsrssstsusvswsxsyszsAsBsCsDsEsFsGsHsIsJsKsLsMsNsOsPsQsRsSsTsUsVsWsXsYsZs0s1s2s3s4s5s6s7s8s9s!s#s$s%s's(s)s*s+s,s-s.s/s:s;s=s?s@s[s]s^s_s`s{s|s}s~satbtctdtetftgthtitjtktltmtntotptqtrtstttutvtwtxtytztAtBtCtDtEtFtGtHtItJtKtLtMtNtOtPtQtRtStTtUtVtWtXtYtZt0t1t2t3t4t5t6t7t8t9t!t#t$t%t't(t)t*t+t,t-t.t/t:t;t=t?t@t[t]t^t_t`t{t|t}t~taubucudueufuguhuiujukulumunuoupuqurusutuuuvuwuxuyuzuAuBuCuDuEuFuGuHuIuJuKuLuMuNuOuPuQuRuSuTuUuVuWuXuYuZu0u1u2u3u4u5u6u7u8u9u!u#u$u%u'u(u)u*u+u,u-u.u/u:u;u=u?u@u[u]u^u_u`u{u|u}u~uavbvcvdvevfvgvhvivjvkvlvmvnvovpvqvrvsvtvuvvvwvxvyvzvAvBvCvDvEvFvGvHvIvJvKvLvMvNvOvPvQvRvSvTvUvVvWvXvYvZv0v1v2v3v4v5v6v7v8v9v!v#v$v%v'v(v)v*v+v,v-v.v/v:v;v=v?v@v[v]v^v_v`v{v|v}v~vawbwcwdwewfwgwhwiwjwkwlwmwnwowpwqwrwswtwuwvwwwxwywzwAwBwCwDwEwFwGwHwIwJwKwLwMwNwOwPwQwRwSwTwUwVwWwXwYwZw0w1w2w3w4w5w6w7w8w9w!w#w$w%w'w(w)w*w+w,w-w.w/w:w;w=w?w@w[w]w^w_w`w{w|w}w~waxbxcxdxexfxgxhxixjxkxlxmxnxoxpxqxrxsxtxuxvxwxxxyxzxAxBxCxDxExFxGxHxIxJxKxLxMxNxOxPxQxRxSxTxUxVxWxXxYxZx0x1x2x3x4x5x6x7x8x9x!x#x$x%x'x(x)x*x+x,x-x.x/x:x;x=x?x@x[x]x^x_x`x{x|x}x~xaybycydyey4e5eTcfygyhyiyjykylymynyoypyqyrysytyuyvywyxyyyzyAyByIgCyDyEyJgFyGyHyIyJyKyLyMyNyOyPyQyRySyTyUyVyWyXyYyZy0y1y2y3y4y5y6y7y8y9y!y#y$y%y'y(y)y*y+y,y-y.y/y:y;y=y?y@y[y]y^y_y`y{y|y}y~yazbzczdzezfzgzhzizjzkzlzmznzozpzqzrzsztzuzvzwzxzyzzzAzBzCzDzEzFzGzHzIzJzKzLzMzNzOzPzQzRzSzTzUzVzWzXzYzZz0z1z2z3z4z5z6z7z8z9z!z#z$z%z'z(z)z*z+z,z-z.z/z:z;z=z?z@z[z]z^z_z`z{z|z}z~zaAbAcAdAeAfAgAhAiAjAkAlAmAnAoApAqArAsAtAuAvAwAxAyAzAAABAUcCAVcDAWcEAXcFA6dGA7dHA8dIA`cJA{cKA|cLA9dMA!dNA#dOA$dPA1cQA2cRA}cSA%dTA~cUA'dVA(dWA)dXA*dYAadZA+d0Abd1Acd2Add3Aed4A,d5A-d6A.d7A/d8A:d9A;d!A=d#A?d$A@d%Afd'Agd(Ahd)Ac *AYc+A,A-A.A/A:A;A=A?A@A[A]A^A_A`Awc{A|A}A~AaBZcbBcBdBeBfBgBhB[diB4g5g6g7g8g9gZh0h1h2h3h4h5h6h7h8h9h!h#h$hjB%h'h(h)h*h+h,h-h.h/hkBlBmBnB:h;h=h]d^d_d`d{doB|d}d~daebepBqBrBsBtBuBvBwBxByBzBABBBCBDBEBFBGBHBIBJBKBLBMBNBOBPBQBRBSBTBUBVBWBXBYBZB0B1B2B3B4B5B6B7B8B9B!B#B$B%B'B(B)B*B+B,B-B.B/B:B;B=B?B@B[B]B^B_B`B{B|B}B~BaCbCcC0cdCeCfCgChCiCjCkClCmCnCoCpCqCrCsCtCuCvCwCxCyCzCACBCCCDCECFCGCHCICJCKCLCMCNCOCPCQCRCSCTCUCVCWCXCYCZC0CUh1C2C3C4C5C6C7C8C9C!C#C$C%C'C(C)C*C+C,C-C.C/C:C;C=C?C@C[C]C^C_C`C{C|C}C~CaDbDcDdDeDfDgDhDiDjDkDlDmDnDoDpDqDrDsDtDuDvDwDxDxfyDzDADBDCDDDEDFDGDHDIDJDKDLDMDNDODPDQDRDSDTDUDVDWDXDYDZD0D1D2D

224 if out_len == 1: 2a Nj@eOj[ePj]eQj8 RjtbSjTjUjubVjWjXjvbYjwbZj0jxb1jyb2jzb3j9 4j! 5j# 6j$ 7j% 8j' 9j( !j) #j* $j+ %j, 'j- (j. )j/ *j: +j; ,j= -j? .j@ /j[ :j] ;j^ =j_ ?j` @j{ [j| ]j^j_j`j{j|j}j~jakbkckdkekfkgkhkikjkkklk^emknkok_epk`eqk{erk|eskAbtk}euk~evkafwkxkykBbzkCbAkDbBkEbCkFbDk(cEk)cFk*cGk+cHk,cIk-cJk.cKk/cLkMkNkOkPkQkRkSkdhTkehUkfhVkWkXkYkZk0k1k2k3k4k5k6k7k8k9k!k#k$k%k'k(k)k*k+k,k-k.k/k:k;k=k?k@k[k]k^k_k`k{k|k}k~kalblcldlelflglhliljlklllmlnlolplqlrlsltlulvlwlxlylzlAlBlClDlElFlGlHlIlJlKlLlMlNlOlPlQlRlSlTlUlVlWlXlYlZl0l1l2l3l4l5l6l7l8l9l!l#l$l%l'l(l)l*l+l,l-l.l/l:l;l=l?l@l[l]l^l_l`l{l|l}l~lambmcmdmemfmgmhmimjmkmlmmmnmompmqmrmsmtmumvmwmxmymzmAmBmCmDmEmFmGm6bHmImJmbfKmLmMmNmOmPmQmRmSmTmUmVmWmXmYmZmcf0mdf1m2m3m4m5m:c6m0d7m8m9mf !mg #m$m%mGb'mob(mHb)m*mef+mjh,mIb-mJb.mKb/mLb:mMb;m=m?m@m[md ]m^m_m`mOePeQeffgfhfh i j k } ~ abbbpbqbrbifjfkflfmfnf{mcb|mNb}mOb~mPbanQbbnRbcndnSbendbfngnofpfhninmhnhciohphdiqhrheishthfiuhgijnLgknqfhilnRemnrfsfiitfnnufjionMgpnvhkiwhxhlivfqnwfrnsntnunvnwnSexnnioipiqirisitiuiviwixiyi7 zie AiHcBiCiDiEiFiGiHiIiJiKiLiMiNiOiPiQiRiSiTiUiViWiXiYiZiFg0i1i2i3i4i5iGg6i7i8i1d9i2d!i;c#i=c$iTbUb?c%i@c'i(i)i*i+iVbWbXbYbZb0bNgynznAnBnCnDnEnFnOgTeGnUeHnInJnKnLnMnNnOnPgPnQnRnSnChTnUnVnDhQgRgWnEhXnYnZn0n1n2n3n4n5n6n7n8n9n!n#n$n%n'nFhGhHh(n)nIhJhKh*nLhMhNh+n,n-n.n/n:n;n=n?n@n[n]n^n_n`n{n|n}n~neb/b:baobocodob ;beo=bfo?bgo@bho[bio]bjo^bko_blo`bmo{bno|boo}bpo~bqoacrobcsocctodcuoecvofcwogcxohcyoiczojcAokcBolcComcDoncEoocFopcGoqcHorcIoJoscKoLoMoNoSgtcucTgUgVgWgXgYgZg0gOhPhce3cfbxcgb4cychb2b6e7e8el %bm n o p q r s t u 'bv w x y 7bz A 8b(bde5cibzcjb6cAckb3b9e!e#eB )bC D E F G H I J K *bL M N O 9bP Q !b+bee7clbBcmb8cCcnb4b$e%e'eR ,bS T U V W X Y Z 0 -b1 2 3 4 #b5 6 $b.b1g[cOoPoQoVeRoSoToUoVoWoXoIcYoWeZo]c0oXe1o2o3o4o5o6o7o8o9o!o#o$o%o'o(o)o*o+o,o-o.o/o:o;o=o?o@o[o]o^o_o`o{o|o}o~oapbpcpdpepfpgphpipjpkplpmpnpopppqprpspYe2g3gZetp^cup0e1evpwpxpypzpApBpvcCpDpEp3dFpGpHpIpJpKpLpMp1bNpOpPpQpRpSpTpUpVpWpXpYpZp0pNc1pOc2pPc3p4p5p6pWhQc7pXhRc8pYhSc9p!p#p$p%p'p(p)p*p+p,p-p.p/p:p;p=p?p@p[p]p^p_p`p{p|p}p~paqbqcqdqeqfqgqhqiqjqkqlqmqnqoqpqqqrqsqtquqvqwqxqyqzqAqBqCqDqEqFqGqHqIqJqKqLqMqNqOqPqQqRqSqTqUqVqWqXqYqZq0q1q2q3q4q5q6q7q8q9q!q#q$q%q'q(q)q*q+q,q-q.q/q:q;q=q?q@q[q]q^q_q`q{q|q}q~qarbrcrdrerfrgrhrirjrkrlrmrnrorprqrrrsrtrurvrwrxryrzrArBrCrDrErFrGrHrIrJrKrLrMrNrOrPrQrRrSrTrUrVrWrXrYrZr0r1r2r3r4r5r6r7r8r9r!r#r$r%r'r(r)r*r+r,r-r.r/r:r;r=r?r@r[r]r^r_r`r{r|r}r~rasbscsdsesfsgshsisjskslsmsnsospsqsrssstsusvswsxsyszsAsBsCsDsEsFsGsHsIsJsKsLsMsNsOsPsQsRsSsTsUsVsWsXsYsZs0s1s2s3s4s5s6s7s8s9s!s#s$s%s's(s)s*s+s,s-s.s/s:s;s=s?s@s[s]s^s_s`s{s|s}s~satbtctdtetftgthtitjtktltmtntotptqtrtstttutvtwtxtytztAtBtCtDtEtFtGtHtItJtKtLtMtNtOtPtQtRtStTtUtVtWtXtYtZt0t1t2t3t4t5t6t7t8t9t!t#t$t%t't(t)t*t+t,t-t.t/t:t;t=t?t@t[t]t^t_t`t{t|t}t~taubucudueufuguhuiujukulumunuoupuqurusutuuuvuwuxuyuzuAuBuCuDuEuFuGuHuIuJuKuLuMuNuOuPuQuRuSuTuUuVuWuXuYuZu0u1u2u3u4u5u6u7u8u9u!u#u$u%u'u(u)u*u+u,u-u.u/u:u;u=u?u@u[u]u^u_u`u{u|u}u~uavbvcvdvevfvgvhvivjvkvlvmvnvovpvqvrvsvtvuvvvwvxvyvzvAvBvCvDvEvFvGvHvIvJvKvLvMvNvOvPvQvRvSvTvUvVvWvXvYvZv0v1v2v3v4v5v6v7v8v9v!v#v$v%v'v(v)v*v+v,v-v.v/v:v;v=v?v@v[v]v^v_v`v{v|v}v~vawbwcwdwewfwgwhwiwjwkwlwmwnwowpwqwrwswtwuwvwwwxwywzwAwBwCwDwEwFwGwHwIwJwKwLwMwNwOwPwQwRwSwTwUwVwWwXwYwZw0w1w2w3w4w5w6w7w8w9w!w#w$w%w'w(w)w*w+w,w-w.w/w:w;w=w?w@w[w]w^w_w`w{w|w}w~waxbxcxdxexfxgxhxixjxkxlxmxnxoxpxqxrxsxtxuxvxwxxxyxzxAxBxCxDxExFxGxHxIxJxKxLxMxNxOxPxQxRxSxTxUxVxWxXxYxZx0x1x2x3x4x5x6x7x8x9x!x#x$x%x'x(x)x*x+x,x-x.x/x:x;x=x?x@x[x]x^x_x`x{x|x}x~xaybycydyey4e5eTcfygyhyiyjykylymynyoypyqyrysytyuyvywyxyyyzyAyByIgCyDyEyJgFyGyHyIyJyKyLyMyNyOyPyQyRySyTyUyVyWyXyYyZy0y1y2y3y4y5y6y7y8y9y!y#y$y%y'y(y)y*y+y,y-y.y/y:y;y=y?y@y[y]y^y_y`y{y|y}y~yazbzczdzezfzgzhzizjzkzlzmznzozpzqzrzsztzuzvzwzxzyzzzAzBzCzDzEzFzGzHzIzJzKzLzMzNzOzPzQzRzSzTzUzVzWzXzYzZz0z1z2z3z4z5z6z7z8z9z!z#z$z%z'z(z)z*z+z,z-z.z/z:z;z=z?z@z[z]z^z_z`z{z|z}z~zaAbAcAdAeAfAgAhAiAjAkAlAmAnAoApAqArAsAtAuAvAwAxAyAzAAABAUcCAVcDAWcEAXcFA6dGA7dHA8dIA`cJA{cKA|cLA9dMA!dNA#dOA$dPA1cQA2cRA}cSA%dTA~cUA'dVA(dWA)dXA*dYAadZA+d0Abd1Acd2Add3Aed4A,d5A-d6A.d7A/d8A:d9A;d!A=d#A?d$A@d%Afd'Agd(Ahd)Ac *AYc+A,A-A.A/A:A;A=A?A@A[A]A^A_A`Awc{A|A}A~AaBZcbBcBdBeBfBgBhB[diB4g5g6g7g8g9gZh0h1h2h3h4h5h6h7h8h9h!h#h$hjB%h'h(h)h*h+h,h-h.h/hkBlBmBnB:h;h=h]d^d_d`d{doB|d}d~daebepBqBrBsBtBuBvBwBxByBzBABBBCBDBEBFBGBHBIBJBKBLBMBNBOBPBQBRBSBTBUBVBWBXBYBZB0B1B2B3B4B5B6B7B8B9B!B#B$B%B'B(B)B*B+B,B-B.B/B:B;B=B?B@B[B]B^B_B`B{B|B}B~BaCbCcC0cdCeCfCgChCiCjCkClCmCnCoCpCqCrCsCtCuCvCwCxCyCzCACBCCCDCECFCGCHCICJCKCLCMCNCOCPCQCRCSCTCUCVCWCXCYCZC0CUh1C2C3C4C5C6C7C8C9C!C#C$C%C'C(C)C*C+C,C-C.C/C:C;C=C?C@C[C]C^C_C`C{C|C}C~CaDbDcDdDeDfDgDhDiDjDkDlDmDnDoDpDqDrDsDtDuDvDwDxDxfyDzDADBDCDDDEDFDGDHDIDJDKDLDMDNDODPDQDRDSDTDUDVDWDXDYDZD0D1D2D

225 return 28 9 ! # $ % ' ( ) * + , - . / : ; = ? @ [ ] ^ _ ` { | :cf g obh i j k } ~ abbbpbqbrbcbdbpfcidieifigiLghiiijiMgkiliSenioipiqirisitiuiviwixiyiziAiHcBiCiDiEiFiGiHiIiJiKiLiMiNiOiPiQiRiSiTiUiViWiXiYiZiFg0i1i2i3i4i5iGg6i7i8i9i!i;c#i=c$iTbUb?c%i@c'i(i)i*i+iVbWbXbYbZb0bTeUeeb/b:bSgtcucTgUgVgWgXgYgZg0gOhPhce3cfbxcgb4cychb2b6e7e8el %bm n o p q r s t u 'bv w x y 7bz A 8b(bde5cibzcjb6cAckb3b9e!e#eB )bC D E F G H I J K *bL M N O 9bP Q !b+bee7clbBcmb8cCcnb4b$e%e'eR ,bS T U V W X Y Z 0 -b1 2 3 4 #b5 6 $b.b1g[cVeIcWe]cXeYe2g3gZe^c0e1e4e5eTc

226 elif out_len == 2: 

227 return result[1] 2a Nj@eOj[ePj]eQj8 RjtbSjTjUjubVjWjXjvbYjwbZj0jxb1jyb2jzb3j9 4j! 5j# 6j$ 7j% 8j' 9j( !j) #j* $j+ %j, 'j- (j. )j/ *j: +j; ,j= -j? .j@ /j[ :j] ;j^ =j_ ?j` @j{ [j| ]j^j_j`j{j|j}j~jakbkckdkekfkgkhkikjkkklk^emknkok_epk`eqk{erk|eskAbtk}euk~evkafwkxkykBbzkCbAkDbBkEbCkFbDk(cEk)cFk*cGk+cHk,cIk-cJk.cKk/cLkMkNkOkPkQkRkSkdhTkehUkfhVkWkXkYkZk0k1k2k3k4k5k6k7k8k9k!k#k$k%k'k(k)k*k+k,k-k.k/k:k;k=k?k@k[k]k^k_k`k{k|k}k~kalblcldlelflglhliljlklllmlnlolplqlrlsltlulvlwlxlylzlAlBlClDlElFlGlHlIlJlKlLlMlNlOlPlQlRlSlTlUlVlWlXlYlZl0l1l2l3l4l5l6l7l8l9l!l#l$l%l'l(l)l*l+l,l-l.l/l:l;l=l?l@l[l]l^l_l`l{l|l}l~lambmcmdmemfmgmhmimjmkmlmmmnmompmqmrmsmtmumvmwmxmymzmAmBmCmDmEmFmGm6bHmImJmbfKmLmMmNmOmPmQmRmSmTmUmVmWmXmYmZmcf0mdf1m2m3m4m5m:c6m0d7m8m9mf !mg #m$m%mGb'mob(mHb)m*mef+mjh,mIb-mJb.mKb/mLb:mMb;m=m?m@m[md ]m^m_m`mOePeQeffgfhfh i j k } ~ abbbpbqbrbifjfkflfmfnf{mcb|mNb}mOb~mPbanQbbnRbcndnSbendbfngnofhninmhnhciohphdiqhrheishthfiuhgijnLgknqfhilnRemnrfsfiitfnnufjionMgpnvhkiwhxhlivfqnwfrnsntnunvnwnSexn7 e 1d2d;c=cTbUb?c@cVbWbXbYbZb0bNgynznAnBnCnDnEnFnOgGnHnInJnKnLnMnNnOnPgPnQnRnSnChTnUnVnDhQgRgWnEhXnYnZn0n1n2n3n4n5n6n7n8n9n!n#n$n%n'nFhGhHh(n)nIhJhKh*nLhMhNh+n,n-n.n/n:n;n=n?n@n[n]n^n_n`n{n|n}n~neb/b:baobocodoeofogohoiojokolomonooopoqorosotouovowoxoyozoAoBoCoDoEoFoGoHoIoJoKoLoMoNoSgtcucTgUgVgWgXgYgZg0gce3cfbxcgb4cychb2b6e7e8el %bm n o p q r s t u 'bv w x y 7bz A 8b(bde5cibzcjb6cAckb3b9e!e#eB )bC D E F G H I J K *bL M N O 9bP Q !b+bee7clbBcmb8cCcnb4b$e%e'eR ,bS T U V W X Y Z 0 -b1 2 3 4 #b5 6 $b.b1gOoPoQoVeRoSoToUoVoWoXoYoZo0o1o2o3o4o5o6o7o8o9o!o#o$o%o'o(o)o*o+o,o-o.o/o:o;o=o?o@o[o]o^o_o`o{o|o}o~oapbpcpdpepfpgphpipjpkplpmpnpopppqprpsp2g3gZetpupvpwpxpypzpApBpCpDpEp3dFpGpHpIpJpKpLpMp1bNpOpPpQpRpSpTpUpVpWpXpYpZp0p1p2p3p4p5p6p7p8p9p!p#p$p%p'p(p)p*p+p,p-p.p/p:p;p=p?p@p[p]p^p_p`p{p|p}p~paqbqcqdqeqfqgqhqiqjqkqlqmqnqoqpqqqrqsqtquqvqwqxqyqzqAqBqCqDqEqFqGqHqIqJqKqLqMqNqOqPqQqRqSqTqUqVqWqXqYqZq0q1q2q3q4q5q6q7q8q9q!q#q$q%q'q(q)q*q+q,q-q.q/q:q;q=q?q@q[q]q^q_q`q{q|q}q~qarbrcrdrerfrgrhrirjrkrlrmrnrorprqrrrsrtrurvrwrxryrzrArBrCrDrErFrGrHrIrJrKrLrMrNrOrPrQrRrSrTrUrVrWrXrYrZr0r1r2r3r4r5r6r7r8r9r!r#r$r%r'r(r)r*r+r,r-r.r/r:r;r=r?r@r[r]r^r_r`r{r|r}r~rasbscsdsesfsgshsisjskslsmsnsospsqsrssstsusvswsxsyszsAsBsCsDsEsFsGsHsIsJsKsLsMsNsOsPsQsRsSsTsUsVsWsXsYsZs0s1s2s3s4s5s6s7s8s9s!s#s$s%s's(s)s*s+s,s-s.s/s:s;s=s?s@s[s]s^s_s`s{s|s}s~satbtctdtetftgthtitjtktltmtntotptqtrtstttutvtwtxtytztAtBtCtDtEtFtGtHtItJtKtLtMtNtOtPtQtRtStTtUtVtWtXtYtZt0t1t2t3t4t5t6t7t8t9t!t#t$t%t't(t)t*t+t,t-t.t/t:t;t=t?t@t[t]t^t_t`t{t|t}t~taubucudueufuguhuiujukulumunuoupuqurusutuuuvuwuxuyuzuAuBuCuDuEuFuGuHuIuJuKuLuMuNuOuPuQuRuSuTuUuVuWuXuYuZu0u1u2u3u4u5u6u7u8u9u!u#u$u%u'u(u)u*u+u,u-u.u/u:u;u=u?u@u[u]u^u_u`u{u|u}u~uavbvcvdvevfvgvhvivjvkvlvmvnvovpvqvrvsvtvuvvvwvxvyvzvAvBvCvDvEvFvGvHvIvJvKvLvMvNvOvPvQvRvSvTvUvVvWvXvYvZv0v1v2v3v4v5v6v7v8v9v!v#v$v%v'v(v)v*v+v,v-v.v/v:v;v=v?v@v[v]v^v_v`v{v|v}v~vawbwcwdwewfwgwhwiwjwkwlwmwnwowpwqwrwswtwuwvwwwxwywzwAwBwCwDwEwFwGwHwIwJwKwLwMwNwOwPwQwRwSwTwUwVwWwXwYwZw0w1w2w3w4w5w6w7w8w9w!w#w$w%w'w(w)w*w+w,w-w.w/w:w;w=w?w@w[w]w^w_w`w{w|w}w~waxbxcxdxexfxgxhxixjxkxlxmxnxoxpxqxrxsxtxuxvxwxxxyxzxAxBxCxDxExFxGxHxIxJxKxLxMxNxOxPxQxRxSxTxUxVxWxXxYxZx0x1x2x3x4x5x6x7x8x9x!x#x$x%x'x(x)x*x+x,x-x.x/x:x;x=x?x@x[x]x^x_x`x{x|x}x~xaybycydyeyTcfygyhyiyjykylymynyoypyqyrysytyuyvywyxyyyzyAyByCyDyEyFyGyHyIyJyKyLyMyNyOyPyQyRySyTyUyVyWyXyYyZy0y1y2y3y4y5y6y7y8y9y!y#y$y%y'y(y)y*y+y,y-y.y/y:y;y=y?y@y[y]y^y_y`y{y|y}y~yazbzczdzezfzgzhzizjzkzlzmznzozpzqzrzsztzuzvzwzxzyzzzAzBzCzDzEzFzGzHzIzJzKzLzMzNzOzPzQzRzSzTzUzVzWzXzYzZz0z1z2z3z4z5z6z7z8z9z!z#z$z%z'z(z)z*z+z,z-z.z/z:z;z=z?z@z[z]z^z_z`z{z|z}z~zaAbAcAdAeAfAgAhAiAjAkAlAmAnAoApAqArAsAtAuAvAwAxAyAzAAABACADAEAFAGAHAIAJAKALAMANAOAPAQARASATAUAVAWAXAYAZA0A1A2A3A4A5A6A7A8A9A!A#A$A%A'A(A)A*A+A,A-A.A/A:A;A=A?A@A[A]A^A_A`A{A|A}A~AaBbBcBdBeBfBgBhBiBjBkBlBmBnBoBpBqBrBsBtBuBvBwBxByBzBABBBCBDBEBFBGBHBIBJBKBLBMBNBOBPBQBRBSBTBUBVBWBXBYBZB0B1B2B3B4B5B6B7B8B9B!B#B$B%B'B(B)B*B+B,B-B.B/B:B;B=B?B@B[B]B^B_B`B{B|B}B~BaCbCcC0cdCeCfCgChCiCjCkClCmCnCoCpCqCrCsCtCuCvCwCxCyCzCACBCCCDCECFCGCHCICJCKCLCMCNCOCPCQCRCSCTCUCVCWCXCYCZC0CUh1C2C3C4C5C6C7C8C9C!C#C$C%C'C(C)C*C+C,C-C.C/C:C;C=C?C@C[C]C^C_C`C{C|C}C~CaDbDcDdDeDfDgDhDiDjDkDlDmDnDoDpDqDrDsDtDuDvDwDxDxfyDzDADBDCDDDEDFDGDHDIDJDKDLDMDNDODPDQDRDSDTDUDVDWDXDYDZD0D1D2D

228 else: 

229 return result[1:] 2a 8 9 ! # $ % ' ( ) * + , - . / : ; = ? @ [ ] ^ _ ` { | cbdbb ;b=b?b@b[b]b^b_b`b{b|b}b~bacbcccdcecfcgchcicjckclcmcncocpcqcrcscvcNcOcPcWhQcXhRcYhScIgJgUcVcWcXc6d7d8d`c{c|c9d!d#d$d1c2c}c%d~c'd(d)d*dad+dbdcddded,d-d.d/d:d;d=d?d@dfdgdhdc YcwcZc[d4g5g6g7g8g9gZh0h1h2h3h4h5h6h7h8h9h!h#h$h%h'h(h)h*h+h,h-h.h/h:h;h=h]d^d_d`d{d|d}d~daebe

230  

231  

232cpdef object check_or_create_options(type cls, object options, str options_description="", bint keep_none=False): 

233 """ 

234 Create the specified options dataclass from a dictionary of options or None. 

235 """ 

236 if options is None: 2)G@e[e]e8 tbfe*Gubge+GvbwbWNxbybzb9 ! # $ % ' ( ) * + , - . / : ; = ? @ [ ] ^ _ ` { | ^e_e`e{e|eAb}e~eafBbCbDbEbFb(c)c*c+c,c-c.c/czfAfBf,G-GidCf:G;GjdDf@G[GkdEf_G`GldFf}G~GmdGfcHdHnd6bbfHfIfJfKfcfdfLfgHhHiH:c0djHkHf g 7DMfGbobHbNfefIbJbKbLbMbd OePeQeffgfhfh i j k } ~ abbbpbqbrbifjfkflfmfnfcbNbObPbQbRbheSbdblHmHofpfnHoHqfpHqHRerfsftfrHsHuftHuHvfvHwfwHxHyHzHAHBHCHDHEHFHGHSeHHIH8DJH9DKHLHMHNHOH!DPH#DQHRHSH7 THe UHyhzh^hHcAhBh_h`hkj9GwNVHWHXHYHZH0H1H2H3H4H5H6H7H8H9H!H#H$H%H'H(H)H*H+H,H-H.H/H:H;H=H?H@H[H1d]H2d^H;c_H=c`HTbUb?c{H@c|H}H~HaIbIVbWbXbYbZb0bXNYNZN0N1N2NzGTeUe3N4N$DDc%D5NcI6NdIeIfIgIhIiIEcjIkIlImI'D(D)DienIoIpIqIrIsItIuIvIwIxIyIzIDGEGAIFGodlj9c*D!c+DGG7NHG8NIGJGKGBILGMGNGOGPGQGRGSGTNUNVN9N!N#N4Deb/b:bpdFcqdrdsdtdudvdwdxdydzdAdBdCdDdEdFdGdHdIdJdKdLdMdNdOdPdQdRdjekeb ;b+M=b,M?b-M@b.M[b/M]b:M^b;M_b=M`b?M{b@M|b[M}b7G~b8Gac]Mbc^Mcc_Mdc`Mec{Mfc|Mgc}Mhc~MicaNjcbNkccNlcdNmceNncyfocfNpcgNqchNrciN,DjNsctcucCIDIce3cfbxcgb4cychb2b6e7e8el %bm n o p q r s t u 'bv w x y 7bz A 8b(bde5cibzcjb6cAckb3b9e!e#eB )bC D E F G H I J K *bL M N O 9bP Q !b+bee7clbBcmb8cCcnb4b$e%e'eR ,bS T U V W X Y Z 0 -b1 2 3 4 #b5 6 $b.bEIFIGI[cHIIIVe-D.DJIKI/D:D;D=DIcWe]cXeLIMINIOIPIQIRISITIUIVIWIXIYIZI0I?D1I@D2I[D3I]D4I^D5I_D6I`D7I{D8I9I!I#I$I%I'I(I)I*I+I,I-I.I/I:I;ITGmjUGVGnjWGYeZe=I?I@I^c0e1e[I]I^I_I|D}DJcKcsbLcMcvc~DaE3dbEcEdEeEfEGcgE1bhEiEjEkElEmEnEoEpEqErEsEtEuENcOcPcvEwExEQcRcScojpjyEzEAEBECEDEEEFEGEHEIEJEKELEMENEOEPEQERESETEUEVEWEXEYEZE0E1E2E3E4E5E6E7E8E9E!E#E$E%E'E(E)E*E+E,E-E.E/E:E;E=E?E@E[E]E^E_E`E{E|E}E~EHg2e3e_c4d`IOf{I|I}IaFPfbFaJbJcFcJdJQfeJfJgJdFRfeFiJSfjJkJTfmJlenJoJXGYGpJqJrJsJtJqjme(euJrjsjvJwJxJneoeUfVfWfXfYfZfyJzJ0fBJ1fCJDJfF2fgFFJ3fGJHJIJhF4fiFKJ5fLJMJ6fOJZGpePJQJRJSJTJUJVJWJXJYJjFqe)eZJtjkF0J1J2J3J4J5J6J7J8J9J!J#J$J%J7f'J(J8f*J+J9flFremFnFoFpFqFrFsFtFuFsevFQhwFxFyF!fteue#f$f%f'f(fSd)f*f#c*e+eTd$cUd+f-J.JVdujvjzFveAFBFCFDFEFFFGFHFIFJFweKFRhLFMFNF,f[J]J-f_J0Gxe`J{J|J}J~JaKbKcKdKeKfKOFye-egKwjPFhKiKjKkK.flKmKnKQF/fRFpK:fqKrK;ftKzeuKvK1G2GwKxKyKzKAKxjAe.eBKyjzjCKDKEKBeCe=f?f@f[f]f^fFKGK_fIK`fJKKKSF{fTFMK|fNKOKPKUF}fVFRK~fSKTKagVK3GDeWKXKYKZK0K1K2K3K4K5KWFEe/e6KAjXF7K8K9K!K#K$K%K'K(K)K*K+K,K-Kbg.K/Kcg;K=KdgYFFeZF0F1F2F3F4F5F6F7FGe8FSh9F!F#FegHeIefggghgigjgWdkglg%c:e;eXd'cYdmg@K[KZdBjCj$FJe%F'F(F)F*F+F,F-F.F/FKe:FTh;F=F?Fng}K~KogbL4GLecLdLeLfLgLhLiLjLkLlLmL@FMe?enLDj[FoLpLqL4e5eTcEj]F^FrLpgsLtLuL_Fqg`F{F|F}FwLxLrgyLzLAL~FsgaGCLDLtgELFLGLbGugcGILJLvgKLLLMLdGwgeGOLPLxgQLRLSLfGyggGULVLzgWLXLYLhGAgiG0L1LBg2L3L4LjGCgkGlG6L7LDg8L9L!LmGEgnGoG$L2M3MUcVc5GWcXc6d7d8d`c{c|c9d!d#d$d1c2c}c%d~c'd(d)d*dad+dbdcddded,d-d.d/d:d;d=d?d@dfdgdhdc YcpG5bqG5d4M5M6M7M8M9M!M#M$M%M'M(M)M*MFjGj%LrGsGtGwcuG'L(LvG6GxGkNlNmNnNZcoNpNqNrNsNtN[d]d^d_d`d{d|d}dWMXMYMZM~d0Maebe1M$N%N'N(N)N*N)L+N,N-N.N/N:N;N=N?NwG@N*L+L[N,L]N0cbiHjIjJjKj`L{L|L}L~LaMcMdMeMfMgMjMLjMjwMxMyMzMAMBMCMDMEMFMGMHMIMJMKMLMMMNMOMPMQMRMSMTMUMVMxf

237 if keep_none: 2)G@e[e]e8 tbfe*Gubge+GvbwbWNxbybzb9 ! # $ % ' ( ) * + , - . / : ; = ? @ [ ] ^ _ ` { | ^e_e`e{e|eAb}e~eafBbCbDbEbFb(c)c*c+c,c-c.c/c,G-G:G;G@G[G_G`G}G~GcHdHbfcfdf:c0df g GbobHbefIbJbKbLbMbd OePeQeffgfhfh i j k } ~ abbbpbqbrbifjfkflfmfnfcbNbObPbQbRbheSbdblHmHofpfnHoHqfpHqHRerfsftfrHsHuftHuHvfvHwfwHxHyHzHAHBHCHDHEHFHGHSe7 e yhzh^hHcAhBh_h`hkj9GwN1d2d;c=cTbUb?c@cVbWbXbYbZb0bXNYNZN0N1N2NTeUe3N4N$DDc%DcIdIeIfIgIhIiIEcjIkIlImI'D(D)DienIoIpIqIrIsItIuIvIwIxIyIzIAIod9c!c8NBIeb/b:bFctcucCIDIce3cfbxcgb4cychb2b6e7e8el %bm n o p q r s t u 'bv w x y 7bz A 8b(bde5cibzcjb6cAckb3b9e!e#eB )bC D E F G H I J K *bL M N O 9bP Q !b+bee7clbBcmb8cCcnb4b$e%e'eR ,bS T U V W X Y Z 0 -b1 2 3 4 #b5 6 $b.bEIFIGI[cHIVe-D.DJIKI/D:D;D=DIcWe]cXe?D1I@D2I[D3I]D4I^D5I_D6I`D7I{D8ImjnjZe=I?I@I^c|D}D~DaE3dbEcEdEeEfEGcgE1bhEiEjEkElEmEnEoEpEqErEsEtEuEQcRcScyEzEAEBECEDEEEFEGEHEIEJEKELEMENEOEPEQERESETEUEVEWEXEYEZE0E1E2E3E4E5E6E7E8E9E!E#E$E%E'E(E)E*E+E,E-E.E/E:E;E=E?E@E[E]E^E_E`E{E|E}E~E_c`I{I|I}IaFbFaJbJcFcJdJeJfJgJdFeFiJjJkJmJlenJoJXGYGpJqJrJsJtJqjme(euJrjsjvJwJxJneoeyJzJBJCJDJfFgFFJGJHJIJhFiFKJLJMJOJZGpePJQJRJSJTJUJVJWJXJYJjFqe)eZJtjkF0J1J2J3J4J5J6J7J8J9J!J#J$J%J'J(J*J+JlFremFnFoFpFqFrFsFtFuFsevFQhwFxFyFteue*e+eTdUd-J.JujvjzFveAFBFCFDFEFFFGFHFIFJFweKFRhLFMFNF[J]J_J0Gxe`J{J|J}J~JaKbKcKdKeKfKOFye-egKwjPFhKiKjKkKlKmKnKQFRFpKqKrKtKzeuKvK1G2GwKxKyKzKAKxjAe.eBKyjzjCKDKEKBeCeFKGKIKJKKKSFTFMKNKOKPKUFVFRKSKTKVK3GDeWKXKYKZK0K1K2K3K4K5KWFEe/e6KAjXF7K8K9K!K#K$K%K'K(K)K*K+K,K-K.K/K;K=KYFFeZF0F1F2F3F4F5F6F7FGe8FSh9F!F#FHeIe:e;eXdYd@K[KBjCj$FJe%F'F(F)F*F+F,F-F.F/FKe:FTh;F=F?F}K~KbL4GLecLdLeLfLgLhLiLjLkLlLmL@FMe?enLDj[FoLpLqLTcEj]F^FrLsLtLuL_F`F{F|F}FwLxLyLzLAL~FaGCLDLELFLGLbGcGILJLKLLLMLdGeGOLPLQLRLSLfGgGULVLWLXLYLhGiG0L1L2L3L4LjGkGlG6L7L8L9L!LmGnGoG$LVcc YcpG5bqG5d%LrGsGtGuG'L(LvGxGZc'N(N)N*N,N-N.N/N=NwG@N[N,L0cbiHjIjJjKj`L{L|L}L~LaMcMdMeMfMgMjMLjMjwMxMyMzMAMBMCMDMEMFMGMHMIMJMKMLMMMNMOMPMQMRMSMTMUMVMxf

238 return options 2f g OePeQeh i j k } ~ abbbmHpfnHqfqHResfrHufuHvHwHyHAHCHEHGHTeUeod9c!cCIDIEIFIGIHIVe-D.DJIKI/D:D;D=DIcWe]cXe?D1I@D2I[D3I]D4I^D5I_D6I`D7I{D8Imjnj=I?I@IGc`IbJdJmJlenJoJXGYGpJqJrJsJtJqjme(euJrjsjvJwJxJBJFJZG0GkKtKzeuKvK1G2GwKxKyKzKAKxjAe.eBKyjzjCKDKEKIKMK3G4GrLxLDLJLPLVL1L7L0cbi`L{L|L}L~LaMcMdMeMfMgMjMLjMjwMxMyMzMAMBMCMDMEMFMGMHMIMJMKMLMMMNMOMPMQMRMSMTMUMVM

239 return cls() 2)G@e[e]e8 tbfe*Gubge+GvbwbWNxbybzb9 ! # $ % ' ( ) * + , - . / : ; = ? @ [ ] ^ _ ` { | ^e_e`e{e|eAb}e~eafBbCbDbEbFb(c)c*c+c,c-c.c/c,G-G:G;G@G[G_G`G}G~GcHdHbfcfdf:c0df g GbobHbefIbJbKbLbMbd OePeQeffgfhfh i j k } ~ abbbpbqbrbifjfkflfmfnfcbNbObPbQbRbheSbdblHofoHpHRerftfsHtHvfwfxHzHBHDHFHSe7 e yhzh^hHcAhBh_h`hkj9GwN1d2d;c=cTbUb?c@cVbWbXbYbZb0bXNYNZN0N1N2N3N4N$DDc%DcIdIeIfIgIhIiIEcjIkIlImI'D(D)DienIoIpIqIrIsItIuIvIwIxIyIzIAIod9c!c8NBIeb/b:bFctcucce3cfbxcgb4cychb2b6e7e8el %bm n o p q r s t u 'bv w x y 7bz A 8b(bde5cibzcjb6cAckb3b9e!e#eB )bC D E F G H I J K *bL M N O 9bP Q !b+bee7clbBcmb8cCcnb4b$e%e'eR ,bS T U V W X Y Z 0 -b1 2 3 4 #b5 6 $b.b[cIcZe^c|D}D~DaE3dbEcEdEeEfEGcgE1bhEiEjEkElEmEnEoEpEqErEsEtEuEQcRcScyEzEAEBECEDEEEFEGEHEIEJEKELEMENEOEPEQERESETEUEVEWEXEYEZE0E1E2E3E4E5E6E7E8E9E!E#E$E%E'E(E)E*E+E,E-E.E/E:E;E=E?E@E[E]E^E_E`E{E|E}E~E_c{I|I}IaFbFaJcFcJeJfJgJdFeFiJjJkJXGYGqjsjneoeyJzJCJDJfFgFGJHJIJhFiFKJLJMJOJZGpePJQJRJSJTJUJVJWJXJYJjFqe)eZJtjkF0J1J2J3J4J5J6J7J8J9J!J#J$J%J'J(J*J+JlFremFnFoFpFqFrFsFtFuFsevFQhwFxFyFteue*e+eTdUd-J.JujvjzFveAFBFCFDFEFFFGFHFIFJFweKFRhLFMFNF[J]J_J0Gxe`J{J|J}J~JaKbKcKdKeKfKOFye-egKwjPFhKiKjKlKmKnKQFRFpKqKrK1G2GxjzjBeCeFKGKJKKKSFTFNKOKPKUFVFRKSKTKVK3GDeWKXKYKZK0K1K2K3K4K5KWFEe/e6KAjXF7K8K9K!K#K$K%K'K(K)K*K+K,K-K.K/K;K=KYFFeZF0F1F2F3F4F5F6F7FGe8FSh9F!F#FHeIe:e;eXdYd@K[KBjCj$FJe%F'F(F)F*F+F,F-F.F/FKe:FTh;F=F?F}K~KbL4GLecLdLeLfLgLhLiLjLkLlLmL@FMe?enLDj[FoLpLqLTcEj]F^FsLtLuL_F`F{F|F}FwLyLzLAL~FaGCLELFLGLbGcGILKLLLMLdGeGOLQLRLSLfGgGULWLXLYLhGiG0L2L3L4LjGkGlG6L8L9L!LmGnGoG$LVcc YcpG5bqG5d%LrGsGtGuG'L(LvGxGZc'N(N)N*N,N-N.N/N=NwG@N[N,L0cbiHjIjJjKjxf

240 elif isinstance(options, cls): 28 tbfeubgevbwbxbybzb9 ! # $ % ' ( ) * + , - . / : ; = ? @ [ ] ^ _ ` { | AbBbCbDbEbFbzfAfBfidCfjdDfkdEfldFfmdGfnd6bHfIfJfKfLfgHhHiH:c0djHkHf g 7DMfGbobHbNfIbJbKbLbMbd h i j k } ~ abbbpbqbrbcbNbObPbQbRbheSbdbHHIH8DJH9DKHLHMHNHOH!DPH#DQHRHSH7 THe UHyhzh^hHcAhBh_h`hkj9GwNVHWHXHYHZH0H1H2H3H4H5H6H7H8H9H!H#H$H%H'H(H)H*H+H,H-H.H/H:H;H=H?H@H[H]H^H_H`HTbUb{H|H}H~HaIbIVbWbXbYbZb0bzG$DDc%D5N6NEc'D(D)DieDGEGFGodlj9c*D!c+DGG7NHGIGJGKGLGMGNGOGPGQGRGSGTNUNVN9N!N#N4DebpdFcqdrdsdtdudvdwdxdydzdAdBdCdDdEdFdGdHdIdJdKdLdMdNdOdPdQdRdjekeb ;b+M=b,M?b-M@b.M[b/M]b:M^b;M_b=M`b?M{b@M|b[M}b7G~b8Gac]Mbc^Mcc_Mdc`Mec{Mfc|Mgc}Mhc~MicaNjcbNkccNlcdNmceNncyfocfNpcgNqchNrciN,DjNsc[cII-D.D/D:D;D=DIc]cLIMINIOIPIQIRISITIUIVIWIXIYIZI0I?D@D[D]D^D_D`D{D9I!I#I$I%I'I(I)I*I+I,I-I.I/I:I;ITGmjUGVGnjWGYe^c0e1e[I]I^I_IJcKcsbLcMcvcGcNcOcPcvEwExEojpjHg2e3e_c4dOfPfQfRfSfTflemeneoeUfVfWfXfYfZf0f1f2f3f4f5f6fpeqe7f8f9frese!fteue#f$f%f'f(fSd)f*f#c$c+fVdvewe,f-fxeye.f/f:f;fzeAeBeCe=f?f@f[f]f^f_f`f{f|f}f~fagDeEebgcgdgFeGeegHeIefggghgigjgWdkglg%c'cmgZdJeKengogLeMe4e5eTcEjpgqgrgsgtgugvgwgxgygzgAgBgCgDgEg2M3MUc5GWcXc6d7d8d`c{c|c9d!d#d$d1c2c}c%d~c'd(d)d*dad+dbdcddded,d-d.d/d:d;d=d?d@dfdgdhd4M5M6M7M8M9M!M#M$M%M'M(M)M*MFjGjwc6GxGkNlNmNnNoNpNqNrNsNtN[d]d^d_d`d{d|d}dWMXMYMZM~d0Maebe1M$N%N)L+N:N;N?NwG*L+L]N

241 return options 28 tbfeubgevbwbxbybzb9 ! # $ % ' ( ) * + , - . / : ; = ? @ [ ] ^ _ ` { | AbBbCbDbEbFbzfAfBfidCfjdDfkdEfldFfmdGfnd6bHfIfJfKfLfgHhHiH:c0djHkHf g 7DMfGbobHbNfIbJbKbLbMbd h i j k } ~ abbbpbqbrbcbNbObPbQbRbheSbdbHHIH8DJH9DKHLHMHNHOH!DPH#DQHRHSH7 THe UH^hHc_h`hwNVHWHXHYHZH0H1H2H3H4H5H6H7H8H9H!H#H$H%H'H(H)H*H+H,H-H.H/H:H;H=H?H@H[H]H^H_H`HTbUb{H|H}H~HaIbIVbWbXbYbZb0b$DDc%D5N6NEc'D(D)DieDGEGFGodlj9c*D!c+DGG7NHGIGJGKGLGMGNGOGPGQGRGSGTNUNVN9N!N#N4DebpdFcqdrdsdtdudvdwdxdydzdAdBdCdDdEdFdGdHdIdJdKdLdMdNdOdPdQdRdjekeb ;b+M=b,M?b-M@b.M[b/M]b:M^b;M_b=M`b?M{b@M|b[M}b7G~b8Gac]Mbc^Mcc_Mdc`Mec{Mfc|Mgc}Mhc~MicaNjcbNkccNlcdNmceNncyfocfNpcgNqchNrciN,DjNsc[cII-D.D/D:D;D=DIc]cLIMINIOIPIQIRISITIUIVIWIXIYIZI0I?D@D[D]D^D_D`D{D9I!I#I$I%I'I(I)I*I+I,I-I.I/I:I;IYe^c0e1e[I]I^I_IJcKcsbLcMcvcGcNcOcPcvEwExEojpjHg2e3e_c4dOfPfQfRfSfTflemeneoeUfVfWfXfYfZf0f1f2f3f4f5f6fpeqe7f8f9frese!fteue#f$f%f'f(fSd)f*f#c$c+fVdvewe,f-fxeye.f/f:f;fzeAeBeCe=f?f@f[f]f^f_f`f{f|f}f~fagDeEebgcgdgFeGeegHeIefggghgigjgWdkglg%c'cmgZdJeKengogLeMe4e5epgqgrgsgtgugvgwgxgygzgAgBgCgDgEg2M3MUc5GWcXc6d7d8d`c{c|c9d!d#d$d1c2c}c%d~c'd(d)d*dad+dbdcddded,d-d.d/d:d;d=d?d@dfdgdhd4M5M6M7M8M9M!M#M$M%M'M(M)M*MFjGj6GxGkNlNmNnNoNpNqNrNsNtN[d]d^d_d`d{d|d}dWMXMYMZM~d0Maebe1M$N%N)L+N:N;N?NwG*L+L]N

242 elif isinstance(options, dict): 2yhzhAhBhkj9GzG4DpdTGmjUGVGnjWGTcEjwc

243 return cls(**options) 2yhzhAhBhkj9GpdTGmjUGVGnjWGTcEjwc

244 else: 

245 raise TypeError( 2zG4D

246 f"The {options_description} must be provided as an object " 2zG4D

247 f"of type {cls.__name__} or as a dict with valid {options_description}. " 2zG4D

248 f"The provided object is '{options}'." 2zG4D

249 ) 

250  

251  

252def _handle_boolean_option(option: bool) -> str: 

253 """ 

254 Convert a boolean option to a string representation. 

255 """ 

256 return "true" if bool(option) else "false" 2b ;b=b?b@b[b]b^b_b`b{b|b}b~bacbcccdcecfcgchcicjckclcmcncocpcqcrcscvcojpj`c{c|c~cddedfdgdhdh6

257  

258  

259def precondition(checker: Callable[..., None], str what="") -> Callable[..., Any]: 

260 """ 

261 A decorator that adds checks to ensure any preconditions are met. 

262  

263 Args: 

264 checker: The function to call to check whether the preconditions are met. It has 

265 the same signature as the wrapped function with the addition of the keyword argument `what`. 

266 what: A string that is passed in to `checker` to provide context information. 

267  

268 Returns: 

269 Callable: A decorator that creates the wrapping. 

270 """ 

271  

272 def outer(wrapped_function: Callable) -> Callable: 2AG

273 """ 

274 A decorator that actually wraps the function for checking preconditions. 

275 """ 

276  

277 @functools.wraps(wrapped_function) 2AG

278 def inner(*args, **kwargs) -> Any: 

279 """ 

280 Check preconditions and if they are met, call the wrapped function. 

281 """ 

282 checker(*args, **kwargs, what=what) 2AG

283 result = wrapped_function(*args, **kwargs) 2AG

284  

285 return result 2AG

286  

287 return inner 2AG

288  

289 return outer 2AG

290  

291  

292def is_sequence(obj: object) -> bool: 

293 """ 

294 Check if the given object is a sequence (list or tuple). 

295 """ 

296 return isinstance(obj, Sequence) 27 e DcEcDGEGFGlj9c!cGGHGIGJGKGLGMGNGOGPGQGRGSGTNUNVNebpdqdrdsdtdudvdwdxdydzdAdBdCdDdEdFdGdHdIdJdKdLdMdNdOdPdQdRd7G8G5G1c2c}cadbdcdi6j6k6l6m6FjGjn6o6p6^5_56G4g5g6g7g8g9gq6r6

297  

298  

299def is_nested_sequence(obj: object) -> bool: 

300 """ 

301 Check if the given object is a nested sequence (list or tuple with atleast one list or tuple element). 

302 """ 

303 return is_sequence(obj) and any(is_sequence(elem) for elem in obj) 21c2c^5_5

304  

305  

306  

307class Transaction: 

308 """ 

309 A context manager for transactional operations with undo capability. 

310  

311 The Transaction class allows you to register undo actions (callbacks) that will be executed 

312 if the transaction is not committed before exiting the context. This is useful for managing 

313 resources or operations that need to be rolled back in case of errors or early exits. 

314  

315 Usage: 

316 with Transaction() as txn: 

317 txn.append(some_cleanup_function, arg1, arg2) 

318 # ... perform operations ... 

319 txn.commit() # Disarm undo actions; nothing will be rolled back on exit 

320  

321 Methods: 

322 append(fn, *args, **kwargs): Register an undo action to be called on rollback. 

323 commit(): Disarm all undo actions; nothing will be rolled back on exit. 

324 """ 

325 def __init__(self) -> None: 

326 self._stack = ExitStack() 2JcKcsbLcMc

327 self._entered = False 2JcKcsbLcMc

328  

329 def __enter__(self): 

330 self._stack.__enter__() 2JcKcsbLcMc

331 self._entered = True 2JcKcsbLcMc

332 return self 2JcKcsbLcMc

333  

334 def __exit__(self, exc_type, exc, tb): 

335 # If exit callbacks remain, they'll run in LIFO order. 

336 self._entered = False 2JcKcsbLcMc

337 return self._stack.__exit__(exc_type, exc, tb) 2JcKcsbLcMc

338  

339 def append(self, fn: Callable[..., Any], /, *args: Any, **kwargs: Any) -> None: 

340 """ 

341 Register an undo action (runs if the with-block exits without commit()). 

342 Values are bound now via partial so late mutations don't bite you. 

343 """ 

344 if not self._entered: 2JcKcsbLcMc

345 raise RuntimeError("Transaction must be entered before append()") 

346 self._stack.callback(partial(fn, *args, **kwargs)) 2JcKcsbLcMc

347  

348 def commit(self) -> None: 

349 """ 

350 Disarm all undo actions. After this, exiting the with-block does nothing. 

351 """ 

352 # pop_all() empties this stack so no callbacks are triggered on exit. 

353 self._stack.pop_all() 2JcKcsbLcMc

354  

355  

356# Track whether we've already warned about fork method 

357_fork_warning_checked = False 

358  

359  

360def reset_fork_warning() -> None: 

361 """Reset the fork warning check flag for testing purposes. 

362  

363 This function is intended for use in tests to allow multiple test runs 

364 to check the warning behavior. 

365 """ 

366 global _fork_warning_checked 

367 _fork_warning_checked = False 2Hg2e3e_c?h4d

368  

369  

370cdef inline tuple _read_fill_ptr(const char* ptr, Py_ssize_t width): 

371 """Extract (value, element_size) from a raw pointer of known width.""" 

372 cdef unsigned int val 

373 if width == 1: 2}i~iajbjcjdjejfjgjhjijjj6Dd 3cfbxcgb4cychbl %bm n o p q r s t u 'bv w x y 7bz A 8b(b5cibzcjb6cAckbB )bC D E F G H I J K *bL M N O 9bP Q !b+b7clbBcmb8cCcnbR ,bS T U V W X Y Z 0 -b1 2 3 4 #b5 6 $b.b

374 val = (<uint8_t*>ptr)[0] 2fbm n o v w ibC D E L M lbS T U 1 2

375 elif width == 2: 

376 val = (<uint16_t*>ptr)[0] 2}iajcjejgjijxcgbp q r x y 7bzcjbF G H N O 9bBcmbV W X 3 4 #b

377 elif width == 4: 

378 val = (<uint32_t*>ptr)[0] 2~ibjdjfjhjjjd ychbl s t u z A 8bAckbB I J K P Q !bCcnbR Y Z 0 5 6 $b

379 else: 

380 raise ValueError(f"value must be 1, 2, or 4 bytes, got {width}") 26D3c4c%b'b(b5c6c)b*b+b7c8c,b-b.b

381 return (val, <unsigned int>width) 2}i~iajbjcjdjejfjgjhjijjjd fbxcgbychbl m n o p q r s t u v w x y 7bz A 8bibzcjbAckbB C D E F G H I J K L M N O 9bP Q !blbBcmbCcnbR S T U V W X Y Z 0 1 2 3 4 #b5 6 $b

382  

383  

384cpdef tuple _parse_fill_value(value): 

385 """Parse a fill/memset value into (raw_value, element_size). 

386  

387 Parameters 

388 ---------- 

389 value : int or buffer-protocol object 

390 - int: Must be in range [0, 256). Treated as 1-byte fill. 

391 - bytes or buffer-protocol: Must be 1, 2, or 4 bytes. 

392  

393 Returns 

394 ------- 

395 tuple of (int, int) 

396 (raw_value, element_size) where element_size is 1, 2, or 4. 

397  

398 Raises 

399 ------ 

400 OverflowError 

401 If int value is outside [0, 256). 

402 TypeError 

403 If value is not an int and does not support the buffer protocol. 

404 ValueError 

405 If value byte length is not 1, 2, or 4. 

406 """ 

407 cdef uint8_t byte_val 

408 cdef Py_buffer buf 

409  

410 if isinstance(value, int): 2(c)c*c+c,c-c.c/c.G/G}i~i=G?Gajbj]G^Gcjdj{G|GejfjaHbHgjhjeHfHijjj6Df g d Neh i j k ce3cfbxcgb4cychb2b6e7e8el %bm n o p q r s t u 'bv w x y 7bz A 8b(bde5cibzcjb6cAckb3b9e!e#eB )bC D E F G H I J K *bL M N O 9bP Q !b+bee7clbBcmb8cCcnb4b$e%e'eR ,bS T U V W X Y Z 0 -b1 2 3 4 #b5 6 $b.b~IhJlJrjAJEJJJNJtj)J,JQhSd,eVdujvj/J:J;J=J?J@JRh^JwjoKsKyjHKLKQKUKAj:K?KShWd=eZdBjCj]K^K_K`K{K|KThaLDjvLBLHLNLTLZL5L#L

411 byte_val = value 2(c)c*c+c,c-c.c/c.G/G=G?G]G^G{G|GaHbHeHfHf g Neh i j k 2b6e7e8e3b9e!e#e4b$e%e'e~IhJlJrjAJEJJJNJtj)J,JQhSd,eVdujvj/J:J;J=J?J@JRh^JwjoKsKyjHKLKQKUKAj:K?KShWd=eZdBjCj]K^K_K`K{K|KThaLDjvLBLHLNLTLZL5L#L

412 return (<unsigned int>byte_val, <unsigned int>1) 2(c)c*c+c,c-c.c/c.G/G=G?G]G^G{G|GaHbHeHfHf g Neh i j k 2b3b4b~IhJlJrjAJEJJJNJtj)J,JQhSd,eVdujvj/J:J;J=J?J@JRh^JwjoKsKyjHKLKQKUKAj:K?KShWd=eZdBjCj]K^K_K`K{K|KThaLDjvLBLHLNLTLZL5L#L

413  

414 if isinstance(value, bytes): 2}i~iajbjcjdjejfjgjhjijjj6Dd ce3cfbxcgb4cychbl %bm n o p q r s t u 'bv w x y 7bz A 8b(bde5cibzcjb6cAckbB )bC D E F G H I J K *bL M N O 9bP Q !b+bee7clbBcmb8cCcnbR ,bS T U V W X Y Z 0 -b1 2 3 4 #b5 6 $b.b

415 return _read_fill_ptr(<const char*><bytes>value, <Py_ssize_t>len(value)) 2}i~iajbjcjdjejfjgjhjijjj6D3cfbxcgb4cychb5cibzcjb6cAckb7clbBcmb8cCcnb

416  

417 if PyObject_GetBuffer(value, &buf, PyBUF_SIMPLE) != 0: 2d cel %bm n o p q r s t u 'bv w x y 7bz A 8b(bdeB )bC D E F G H I J K *bL M N O 9bP Q !b+beeR ,bS T U V W X Y Z 0 -b1 2 3 4 #b5 6 $b.b

418 raise TypeError( 

419 f"value must be an int or support the buffer protocol, " 

420 f"got {type(value).__name__}" 

421 ) 

422 try: 2d l %bm n o p q r s t u 'bv w x y 7bz A 8b(bB )bC D E F G H I J K *bL M N O 9bP Q !b+bR ,bS T U V W X Y Z 0 -b1 2 3 4 #b5 6 $b.b

423 return _read_fill_ptr(<const char*>buf.buf, buf.len) 2d l %bm n o p q r s t u 'bv w x y 7bz A 8b(bB )bC D E F G H I J K *bL M N O 9bP Q !b+bR ,bS T U V W X Y Z 0 -b1 2 3 4 #b5 6 $b.b

424 finally: 

425 PyBuffer_Release(&buf) 

426  

427  

428def check_multiprocessing_start_method() -> None: 

429 """Check if multiprocessing start method is 'fork' and warn if so.""" 

430 global _fork_warning_checked 

431 if _fork_warning_checked: 2xNyN8D9DzNANBNCN!D#D7 e yhzh^hHcAhBh_h`hDNEN`5FN{5|5GN}5~5HNa6b6INc6d6JNe6f6Fgg6KNLNMNGgNNON1d2d;c=cTbUb?c@cPNQNRNSNVbWbXbYbZb0bHg2e3e_c?h4d

432 return 2xNyN8D9DzNANBNCN!D#D7 e yhzh^hHcAhBh_h`hDNEN`5FN{5|5GN}5~5HNa6b6INc6d6JNe6f6Fgg6KNLNMNGgNNON1d2d;c=cTbUb?c@cPNQNRNSNVbWbXbYbZb0b4d

433 _fork_warning_checked = True 2e Hg2e3e_c?h4d

434  

435 # Common warning message parts 

436 common_message = ( 

437 "CUDA does not support. Forked subprocesses exhibit undefined behavior, " 2e Hg2e3e_c?h4d

438 "including failure to initialize CUDA contexts and devices. Set the start method " 

439 "to 'spawn' before creating processes that use CUDA. " 

440 "Use: multiprocessing.set_start_method('spawn')" 

441 ) 

442  

443 try: 2e Hg2e3e_c?h4d

444 start_method = multiprocessing.get_start_method() 2e Hg2e3e_c?h4d

445 if start_method == "fork": 2e Hg2e3e_c4d

446 message = f"multiprocessing start method is 'fork', which {common_message}" 22e3e_c4d

447 warnings.warn(message, UserWarning, stacklevel=3) 22e3e_c4d

448 except RuntimeError: 2?h

449 # get_start_method() can raise RuntimeError if start method hasn't been set 

450 # In this case, default is 'fork' on Linux, so we should warn 

451 if platform.system() == "Linux": 2?h

452 message = ( 

453 f"multiprocessing start method is not set and defaults to 'fork' on Linux, " 2?h

454 f"which {common_message}" 

455 ) 

456 warnings.warn(message, UserWarning, stacklevel=3) 2?h