
    sh,                     ~    S SK Jr  S SKrS SKJrJr  S SKrS SKr " S S\5      r	\
S:X  a  \R                  " 5         gg)    )cudaN)skip_on_cudasimCUDATestCasec                   8   \ rS rSr\R
                  " \" \R                  5      S:  S5      S 5       r	\
" S5      S 5       r\R
                  " \" \R                  5      S:  S5      S 5       r\R
                  " \" \R                  5      S:  S5      S 5       rS	rg
)TestMultiGPUContext      zneed more than 1 gpusc                 \   [         R                  " S5      S 5       nS nSn[        R                  " U[        R                  S9n[        R                  " U[        R                  S9n[         R
                  S      USU4   " XE5        S S S 5        U" XE5        USU4   " XE5        U" XE5        [         R
                  S      [        R                  " U[        R                  S9n[        R                  " U[        R                  S9nUSU4   " Xg5        [         R
                  S      [        R                  " U[        R                  S9n[        R                  " U[        R                  S9n	USU4   " X5        S S S 5        S S S 5        U" WW5        U" WW	5        [        R                  " U[        R                  S9n[        R                  " U[        R                  S9nUSU4   " XE5        U" XE5        g ! , (       d  f       GNx= f! , (       d  f       N= f! , (       d  f       N= f)Nzvoid(float64[:], float64[:])c                 b    [         R                  " S5      nX!R                  :  a
  X   S-   X'   g g N   r   gridsize)inpoutis      y/Users/tiagomarins/Projetos/claudeai/copy_bank/venv/lib/python3.13/site-packages/numba/cuda/tests/cudapy/test_multigpu.pycopy_plus_1>TestMultiGPUContext.test_multigpu_context.<locals>.copy_plus_1   s*    		!A88|!     c                 J    [         R                  R                  U S-   U5        g r   )nptestingassert_equal)r   r   s     r   check8TestMultiGPUContext.test_multigpu_context.<locals>.check   s    JJ##C!GS1r       dtyper   r   )r   jitr   arangefloat64gpus)
selfr   r   NABA0B0A1B1s
             r   test_multigpu_context)TestMultiGPUContext.test_multigpu_context	   s   	0	1	$ 
2	$
	2 IIarzz*IIarzz*YYq\1a#  	aAqD!aYYq\1BJJ/B1BJJ/B1b%1YYq

3YYq

3AqD!")   	b"b"IIarzz*IIarzz*AqD!a1 \  \s2   ;G:A'H(AH<H:
H	
H	H
H+z+Simulator does not support multiple threadsc           
         S n[         R                  " [        R                  " S5      5      nSnS /U-  n[	        U5       Vs/ s H3  n[
        R                  " U[         R                  R                  X$U4S9PM5     nnU H  nUR                  5         M     U H  nUR                  5         M     U H+  n[        U[        5      (       a  UeU R                  U5        M-     g s  snf )Nc                      U    UR                  5       nS S S 5        [        R                  " W[        R                  " S5      :H  5      X#'   g ! , (       d  f       N>= f! [         a  nXRU'    S nAg S nAff = f)N
   )copy_to_hostr   allr"   	Exception)gpudAresultsridxarres         r   work4TestMultiGPUContext.test_multithreaded.<locals>.work4   sZ    =//+C  !#sbiim'; < S  " !"s-   A AA 
AA 
A6(A11A6r1   )targetargs)r   	to_devicer   r"   range	threadingThreadr$   currentstartjoin
isinstanceBaseException
assertTrue)	r%   r;   r6   nthreadsr7   r   threadsthrs	            r   test_multithreaded&TestMultiGPUContext.test_multithreaded2   s    		= ^^BIIbM*&8# "(O-+q ##4tyy7H7H79A7G H+ 	 - BHHJ  BGGI  A!]++"	 -s   :C#c                    [         R                  S 5       n[        R                  " S[        R                  S9n[         R
                  S      [         R                  " U5      nS S S 5        [         R
                  S      [         R                  " U5      nS S S 5        [         R
                  S      US   " WS5        S S S 5        [         R
                  S      US   " WS5        S S S 5        [         R
                  S      [        R                  R                  WR                  5       US-   5        S S S 5        [         R
                  S      [        R                  R                  WR                  5       US-   5        S S S 5        g ! , (       d  f       GN,= f! , (       d  f       GN= f! , (       d  f       N= f! , (       d  f       N= f! , (       d  f       N= f! , (       d  f       g = f)Nc                 h    [         R                  " S5      nX R                  :  a  X==   U-  ss'   g g r   r   )r9   valr   s      r   vector_add_scalar@TestMultiGPUContext.test_with_context.<locals>.vector_add_scalarU   s(    		!A88|# r   r1   r   r   r   )r   r1   r	   )
r   r!   r   r"   float32r$   r?   r   r   r2   )r%   rR   hostarrarr1arr2s        r   test_with_context%TestMultiGPUContext.test_with_contextR   sH    
	 
	
 ))Bbjj1YYq\>>'*D  YYq\>>'*D  YYq\e$T1-  YYq\e$T1-  YYq\JJ##D$5$5$7'A+G  YYq\JJ##D$5$5$7'A+G \ \ \ \ \ \ \sH   F	?F1F-F>2G2G 	
F
F*-
F;>
G
G 
G.c                 *   [         R                  S      [         R                  " 5       nUR                  S5      (       d  U R	                  S5        S S S 5        [
        R                  " S[
        R                  S9n[         R                  S      [         R                  " U5      nS S S 5        [         R                  S      [         R                  " [
        R                  " U5      5      nS S S 5        [         R                  S      WR                  W5        [
        R                  R                  UR                  5       U5        S S S 5        g ! , (       d  f       GN= f! , (       d  f       N= f! , (       d  f       N= f! , (       d  f       g = f)Nr   r   z!Peer access between GPUs disabledr1   r   )r   r$   current_contextcan_access_peerskipTestr   r"   rT   r?   
zeros_likecopy_to_devicer   r   r2   )r%   ctxrU   rV   rW   s        r   test_with_context_peer_copy/TestMultiGPUContext.test_with_context_peer_copyn   s    
 YYq\&&(C&&q))AB  ))Bbjj1 YYq\>>'*D  YYq\>>"--"89D  YYq\% JJ##D$5$5$7A \! \ \ \ \s0   =EE"+E3A F
E"
E03
F
F N)__name__
__module____qualname____firstlineno__unittestskipIflenr   r$   r-   r   rM   rX   ra   __static_attributes__rc   r   r   r   r      s    __S^a')@A& B&P BC# D#> __S^a')@AH BH6 __S^a')@AB BBr   r   __main__)numbar   numpyr   numba.cuda.testingr   r   rA   rh   r   rd   mainrc   r   r   <module>rq      s=      <  @B, @BF zMMO r   