U
    ?hc  ã                   @   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rˆe ¡  dS )é    N)Úcuda)ÚunittestÚCUDATestCaseÚskip_on_cudasimc                    s   t  ˆ ¡‡ fdd„ƒ}|S )Nc                     s6   t  ¡ }| d¡ z| ˆ | |Ž¡W ¢S | ¡  X d S )NT)ÚasyncioZnew_event_loopZ	set_debugÚcloseZrun_until_complete)ÚargsÚkwdsÚloop©Úf© úW/var/www/html/venv/lib/python3.8/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   )Z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   t t dœ‡fdd„‰ ddddg}‡ fdd	„|D ƒ}tj|Ž I d H }ˆ t ||¡¡ d S )
N)Úvalue_inÚreturnc                 “   sh   t  ¡ }t  d¡t  d¡ }}| |d d …< t j||d}|j||d | ¡ I d H }ˆ  ||¡ | ¡ S )Né   ©r   )r   r   Zpinned_arrayZ	to_deviceZcopy_to_hostr&   ÚassertEqualZmean)r(   r   Zh_srcZh_dstZd_aryZ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   Zcreate_task)Ú.0Úv)r-   r   r   Ú
<listcomp>=   s     z6TestCudaStream.test_parallel_tasks.<locals>.<listcomp>)Úfloatr   Úgatherr   ÚnpZallclose)r    Z	values_inÚtasksZ
values_outr   )r-   r    r   Útest_parallel_tasks0   s
    
z"TestCudaStream.test_parallel_tasksc                 ƒ   sH   t  ¡ ‰ ‡ fdd„tdƒD ƒ}tj|Ž I d H }|D ]}|  |ˆ ¡ q2d S )Nc                    s   g | ]}ˆ   ¡ ‘qS r   ©r&   ©r2   Ú_r+   r   r   r4   D   s     z;TestCudaStream.test_multiple_async_done.<locals>.<listcomp>r1   )r   r   Úranger   r6   r,   )r    Údone_awsÚdoneÚdr   r+   r   Útest_multiple_async_doneA   s
    z'TestCudaStream.test_multiple_async_donec                 Ã   sH   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   r;   r   r   r   r4   K   s     zLTestCudaStream.test_multiple_async_done_multiple_streams.<locals>.<listcomp>r1   c                 S   s   g | ]}|  ¡ ‘qS r   r:   )r2   r   r   r   r   r4   L   s     )r=   r   r6   ZassertSetEqualr   )r    Ústreamsr>   r?   r   r   r   Ú)test_multiple_async_done_multiple_streamsI   s    z8TestCudaStream.test_multiple_async_done_multiple_streamsc                 Ã   sL   t  ¡ }| ¡ | ¡  }}| ¡  |I d H  |  | ¡ ¡ |  | ¡ ¡ d S r   )r   r   r&   Úcancelr   Z	cancelledr?   )r    r   Zdone1Zdone2r   r   r   Útest_cancelled_futureR   s    
z$TestCudaStream.test_cancelled_futureN)Ú__name__Ú
__module__Ú__qualname__r"   r%   r   r'   r9   rA   rC   rE   r   r   r   r   r      s   	



r   c                   @   s   e Zd Zejedd„ ƒƒZdS )ÚTestFailingStreamc              	   Ã   sr   t  ¡ }| d¡}| d¡}t  ¡ }|jdd|d ¡  | ¡ }|  t	¡ |I d H  W 5 Q R X |  
| ¡ ¡ d S )Nz
            .version 6.5
            .target sm_30
            .address_size 64
            .visible .entry failing_kernel() { trap; }
        Úfailing_kernel)r.   r+   )r   Zcurrent_contextZcreate_module_ptxZget_functionr   Ú	configureÚ__call__r&   ZassertRaisesÚ	ExceptionZassertIsNotNoneÚ	exception)r    ÚctxÚmodulerJ   r   r?   r   r   r   Útest_failed_streame   s    

z$TestFailingStream.test_failed_streamN)rF   rG   rH   r   Úskipr   rQ   r   r   r   r   rI   \   s   	rI   Ú__main__)r   r   r   Únumpyr7   Znumbar   Znumba.cuda.testingr   r   r   r   r   rI   rF   Úmainr   r   r   r   Ú<module>   s   F