o
    ic                     @   s   d dl Z d dlZd dlZd dlZd dlmZ d dlmZm	Z	m
Z
 dd Ze
dG dd de	Ze
dG d	d
 d
e	ZedkrFe  dS dS )    N)cuda)unittestCUDATestCaseskip_on_cudasimc                    s   t   fdd}|S )Nc                     s<   t  }|d z| | i |W |  S |  w )NT)asyncionew_event_loop	set_debugrun_until_completeclose)argskwdsloopf c/home/ubuntu/transcripts/venv/lib/python3.10/site-packages/numba/cuda/tests/cudadrv/test_streams.pyrunner
   s
   
z!with_asyncio_loop.<locals>.runner)	functoolswraps)r   r   r   r   r   with_asyncio_loop	   s   r   z,CUDA Driver API unsupported in the simulatorc                   @   sX   e Zd Zdd Zdd Zedd Zedd Zed	d
 Zedd Z	edd Z
dS )TestCudaStreamc                 C   s8   dd }t  }t }||| | |d d S )Nc                 S   s   |   d S N)set)streamstatuseventr   r   r   callback   s   z2TestCudaStream.test_add_callback.<locals>.callback      ?)r   r   	threadingEventadd_callback
assertTruewait)selfr   r   callback_eventr   r   r   test_add_callback   s
   z TestCudaStream.test_add_callbackc                    s<   t    fdd}t }||  d d S )Nc                    s    |    d S r   )assertIsNoner   )r   r   argr$   r#   r   r   r   #   s   
zCTestCudaStream.test_add_callback_with_default_arg.<locals>.callbackr   )r   r   r   r   r    r!   r"   )r#   r   r   r   r(   r   "test_add_callback_with_default_arg    s
   
z1TestCudaStream.test_add_callback_with_default_argc                    s   t  }| I d H  d S r   )r   r   
async_done)r#   r   r   r   r   test_async_done+   s   zTestCudaStream.test_async_donec                    sX   dt dt ffdd g d} fdd|D }tj| I d H }t|| d S )Nvalue_inreturnc                    sj   t  }t dt d}}| |d d < t j||d}|j||d | I d H } || | S )N   r   )r   r   pinned_array	to_devicecopy_to_hostr*   assertEqualmean)r,   r   h_srch_dstd_arydone_result)r#   r   r   async_cuda_fn2   s   z9TestCudaStream.test_parallel_tasks.<locals>.async_cuda_fn)            c                    s   g | ]	}t  |qS r   )r   create_task).0v)r9   r   r   
<listcomp>=   s    z6TestCudaStream.test_parallel_tasks.<locals>.<listcomp>)floatr   gatherr!   npallclose)r#   	values_intasks
values_outr   )r9   r#   r   test_parallel_tasks0   s   
z"TestCudaStream.test_parallel_tasksc                    sJ   t    fddtdD }tj| I d H }|D ]}| |  qd S )Nc                    s   g | ]}   qS r   r*   r?   _r/   r   r   rA   D       z;TestCudaStream.test_multiple_async_done.<locals>.<listcomp>r=   )r   r   ranger   rC   r3   )r#   done_awsdonedr   r/   r   test_multiple_async_doneA   s   z'TestCudaStream.test_multiple_async_donec                    sJ   dd t dD }dd |D }tj| I d H }| t|t| d S )Nc                 S   s   g | ]}t  qS r   )r   r   rK   r   r   r   rA   K   rM   zLTestCudaStream.test_multiple_async_done_multiple_streams.<locals>.<listcomp>r=   c                 S   s   g | ]}|  qS r   rJ   )r?   r   r   r   r   rA   L   rM   )rN   r   rC   assertSetEqualr   )r#   streamsrO   rP   r   r   r   )test_multiple_async_done_multiple_streamsI   s
   z8TestCudaStream.test_multiple_async_done_multiple_streamsc                    sN   t  }| | }}|  |I d H  | |  | |  d S r   )r   r   r*   cancelr!   	cancelledrP   )r#   r   done1done2r   r   r   test_cancelled_futureR   s   
z$TestCudaStream.test_cancelled_futureN)__name__
__module____qualname__r%   r)   r   r+   rI   rR   rU   rZ   r   r   r   r   r      s    	



r   c                   @   s   e Zd Zejedd ZdS )TestFailingStreamc                    s   t  }|d}|d}t  }|jdd|d  | }| t	 |I d H  W d    n1 s6w   Y  | 
|  d S )Nz
            .version 6.5
            .target sm_30
            .address_size 64
            .visible .entry failing_kernel() { trap; }
        failing_kernel)r:   r/   )r   current_contextcreate_module_ptxget_functionr   	configure__call__r*   assertRaises	ExceptionassertIsNotNone	exception)r#   ctxmoduler_   r   rP   r   r   r   test_failed_streame   s   

z$TestFailingStream.test_failed_streamN)r[   r\   r]   r   skipr   rk   r   r   r   r   r^   \   s    	r^   __main__)r   r   r   numpyrD   numbar   numba.cuda.testingr   r   r   r   r   r^   r[   mainr   r   r   r   <module>   s    F