o
    i,                     @   s\   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 deZ	e
dkr,e  dS dS )    )cudaN)skip_on_cudasimCUDATestCasec                   @   sv   e Zd Zeeejdk ddd Ze	ddd Z
eeejdk ddd	 Zeeejdk dd
d ZdS )TestMultiGPUContext   zneed more than 1 gpusc           
   	   C   s  t ddd }dd }d}tj|tjd}tj|tjd}t jd  |d	|f || W d    n1 s8w   Y  ||| |d	|f || ||| t jd O tj|tjd}tj|tjd}|d	|f || t jd	 ! tj|tjd}tj|tjd}	|d	|f ||	 W d    n1 sw   Y  W d    n1 sw   Y  ||| |||	 tj|tjd}tj|tjd}|d	|f || ||| d S )
Nzvoid(float64[:], float64[:])c                 S   s,   t d}||jk r| | d ||< d S d S N   r   gridsize)inpouti r   b/home/ubuntu/veenaModal/venv/lib/python3.10/site-packages/numba/cuda/tests/cudapy/test_multigpu.pycopy_plus_1      

z>TestMultiGPUContext.test_multigpu_context.<locals>.copy_plus_1c                 S   s   t j| d | d S r   )nptestingassert_equal)r   r   r   r   r   check   s   z8TestMultiGPUContext.test_multigpu_context.<locals>.check    dtyper   r   )r   jitr   arangefloat64gpus)
selfr   r   NABA0B0A1B1r   r   r   test_multigpu_context	   s:   





z)TestMultiGPUContext.test_multigpu_contextz+Simulator does not support multiple threadsc                    s   dd t td d}d g|  fddt|D }|D ]}|  q"|D ]}|  q+D ]}t|tr=|| 	| q4d S )Nc              
   S   s|   z|  |  }W d    n1 sw   Y  W n ty/ } z|||< W Y d }~d S d }~ww t|tdk||< d S )N
   )copy_to_host	Exceptionr   allr   )gpudAresultsridxarrer   r   r   work4   s   
z4TestMultiGPUContext.test_multithreaded.<locals>.workr'   c                    s&   g | ]}t jtjj |fd qS ))targetargs)	threadingThreadr   r   current).0r   r,   r-   r1   r   r   
<listcomp>C   s
    z:TestMultiGPUContext.test_multithreaded.<locals>.<listcomp>)
r   	to_devicer   r   rangestartjoin
isinstanceBaseException
assertTrue)r   nthreadsthreadsthrr   r8   r   test_multithreaded2   s    



z&TestMultiGPUContext.test_multithreadedc                 C   s|  t jdd }tjdtjd}t jd  t |}W d    n1 s$w   Y  t jd  t |}W d    n1 s>w   Y  t jd  |d |d W d    n1 sZw   Y  t jd  |d |d W d    n1 svw   Y  t jd  tj|	 |d  W d    n1 sw   Y  t jd  tj|	 |d  W d    d S 1 sw   Y  d S )	Nc                 S   s,   t d}|| jk r| |  |7  < d S d S r   r	   )r/   valr   r   r   r   vector_add_scalarU   r   z@TestMultiGPUContext.test_with_context.<locals>.vector_add_scalarr'   r   r   r   )r   r'   r   )
r   r   r   r   float32r   r:   r   r   r(   )r   rG   hostarrarr1arr2r   r   r   test_with_contextR   s*   
"z%TestMultiGPUContext.test_with_contextc                 C   s  t jd  t  }|ds| d W d    n1 sw   Y  tjdtjd}t jd  t |}W d    n1 s@w   Y  t jd  t t	|}W d    n1 s]w   Y  t jd  |
| tj| | W d    d S 1 sw   Y  d S )Nr   r   z!Peer access between GPUs disabledr'   r   )r   r   current_contextcan_access_peerskipTestr   r   rH   r:   
zeros_likecopy_to_devicer   r   r(   )r   ctxrI   rJ   rK   r   r   r   test_with_context_peer_copyn   s"   


"z/TestMultiGPUContext.test_with_context_peer_copyN)__name__
__module____qualname__unittestskipIflenr   r   r&   r   rE   rL   rS   r   r   r   r   r      s    
(

r   __main__)numbar   numpyr   numba.cuda.testingr   r   r4   rW   r   rT   mainr   r   r   r   <module>   s     