
    xKg,                     |    d dl mZ d dlZd dlmZmZ d dlZd dlZ G d de      Z	e
dk(  r ej                          yy)    )cudaN)skip_on_cudasimCUDATestCasec                   0   e Zd Z ej                   eej                        dk  d      d        Z e	d      d        Z
 ej                   eej                        dk  d      d        Z ej                   eej                        dk  d      d        Zy)	TestMultiGPUContext   zneed more than 1 gpusc                 l   t        j                  d      d        }d }d}t        j                  |t        j                        }t        j                  |t        j                        }t         j
                  d   5   |d|f   ||       d d d         |||        |d|f   ||        |||       t         j
                  d   5  t        j                  |t        j                        }t        j                  |t        j                        } |d|f   ||       t         j
                  d   5  t        j                  |t        j                        }t        j                  |t        j                        }	 |d|f   ||	       d d d        d d d         |        |	       t        j                  |t        j                        }t        j                  |t        j                        } |d|f   ||        |||       y # 1 sw Y   xY w# 1 sw Y   xY w# 1 sw Y   xY w)Nzvoid(float64[:], float64[:])c                 d    t        j                  d      }||j                  k  r| |   dz   ||<   y y N   r   gridsize)inpoutis      i/home/alanp/www/video.onchill/myenv/lib/python3.12/site-packages/numba/cuda/tests/cudapy/test_multigpu.pycopy_plus_1z>TestMultiGPUContext.test_multigpu_context.<locals>.copy_plus_1   s0    		!A388|Q!A     c                 J    t         j                  j                  | dz   |       y r   )nptestingassert_equal)r   r   s     r   checkz8TestMultiGPUContext.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_contextz)TestMultiGPUContext.test_multigpu_context	   s   	0	1	$ 
2	$
	2 IIarzz*IIarzz*YYq\K1a#  	aAqD!QaYYq\1BJJ/B1BJJ/BK1b"%1YYq

3YYq

3!AqD!"b)   	b"b"IIarzz*IIarzz*AqD!Qa1 \  \s2   >HA,H*4AHH*HH'	#H**H3z+Simulator does not support multiple threadsc           
         d }t        j                  t        j                  d            }d}d g|z  }t	        |      D cg c]5  }t        j                  |t         j                  j                  |||f      7 }}|D ]  }|j                           |D ]  }|j                           |D ]%  }t        |t              r|| j                  |       ' y c c}w )Nc                     	 | 5  |j                         }d d d        t        j                  t        j                  d      k(        ||<   y # 1 sw Y   8xY w# t        $ r}|||<   Y d }~y d }~ww xY w)N
   )copy_to_hostr   allr   	Exception)gpudAresultsridxarres         r   workz4TestMultiGPUContext.test_multithreaded.<locals>.work4   s^    =//+C  !#sbiim'; < S  " !"s-   A AA AA 	A0!A++A0r-   )targetargs)r   	to_devicer   r   range	threadingThreadr!   currentstartjoin
isinstanceBaseException
assertTrue)	r"   r7   r2   nthreadsr3   r   threadsthrs	            r   test_multithreadedz&TestMultiGPUContext.test_multithreaded2   s    		= ^^BIIbM*&8# "(O-+q ##4tyy7H7H797A7G H+ 	 - BHHJ  BGGI  A!]+"	 -s   :Cc                    t         j                  d        }t        j                  dt        j                        }t         j
                  d   5  t        j                  |      }d d d        t         j
                  d   5  t        j                  |      }d d d        t         j
                  d   5   |d   d       d d d        t         j
                  d   5   |d   d       d d d        t         j
                  d   5  t        j                  j                  j                         |dz          d d d        t         j
                  d   5  t        j                  j                  j                         |dz          d d d        y # 1 sw Y   &xY w# 1 sw Y   xY w# 1 sw Y   xY w# 1 sw Y   xY w# 1 sw Y   xY w# 1 sw Y   y xY w)Nc                 h    t        j                  d      }|| j                  k  r| |xx   |z  cc<   y y r   r   )r5   valr   s      r   vector_add_scalarz@TestMultiGPUContext.test_with_context.<locals>.vector_add_scalarU   s,    		!A388|A# r   r-   r   r   r   )r   r-   r   )
r   r   r   r   float32r!   r:   r   r   r.   )r"   rL   hostarrarr1arr2s        r   test_with_contextz%TestMultiGPUContext.test_with_contextR   sN    
	 
	
 ))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/2F;2G	FF #F,/F8;GGc                    t         j                  d   5  t        j                         }|j                  d      s| j	                  d       d d d        t        j                  dt
        j                        }t         j                  d   5  t        j                  |      }d d d        t         j                  d   5  t        j                  t        j                  |            }d d d        t         j                  d   5  j                         t
        j                  j                  |j                         |       d d d        y # 1 sw Y   xY w# 1 sw Y   xY w# 1 sw Y   }xY w# 1 sw Y   y xY w)Nr   r   z!Peer access between GPUs disabledr-   r   )r   r!   current_contextcan_access_peerskipTestr   r   rM   r:   
zeros_likecopy_to_devicer   r   r.   )r"   ctxrN   rO   rP   s        r   test_with_context_peer_copyz/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   7E	E<)E! A E-	EE!E*-E6N)__name__
__module____qualname__unittestskipIflenr   r!   r*   r   rH   rQ   rY    r   r   r   r      s    X__S^a')@A& B&P BC# D#> X__S^a')@AH BH6 X__S^a')@AB BBr   r   __main__)numbar   numpyr   numba.cuda.testingr   r   r<   r]   r   rZ   mainr`   r   r   <module>rf      s?      <  @B, @BF zHMMO r   