§
    ƒÛgc  ã                   óö   — 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„ Z e
d¦  «         G d„ de	¦  «        ¦   «         Z e
d¦  «         G d„ d	e	¦  «        ¦   «         Zed
k    r ej        ¦   «          dS dS )é    N)Úcuda)ÚunittestÚCUDATestCaseÚskip_on_cudasimc                 óF   ‡ — t          j        ‰ ¦  «        ˆ fd„¦   «         }|S )Nc                  óæ   •— t          j        ¦   «         }|                     d¦  «         	 |                      ‰| i |¤Ž¦  «        |                     ¦   «          S # |                     ¦   «          w xY w)NT)ÚasyncioÚnew_event_loopÚ	set_debugÚrun_until_completeÚclose)ÚargsÚkwdsÚloopÚfs      €ú[/granary/edev/quantum/lib/python3.11/site-packages/numba/cuda/tests/cudadrv/test_streams.pyÚrunnerz!with_asyncio_loop.<locals>.runner
   si   ø€ åÔ%Ñ'Ô'ˆØ�Š�tÑÔÐð	Ø×*Ò*¨1¨1¨dÐ+;°dÐ+;Ð+;Ñ<Ô<à�JŠJ‰LŒLˆLˆLøˆD�JŠJ‰LŒLˆLˆLøøøs   «A ÁA0)Ú	functoolsÚwraps)r   r   s   ` r   Úwith_asyncio_loopr   	   s8   ø€ Ý„_�QÑÔðð ð ð ñ Ôðð €Mó    z,CUDA Driver API unsupported in the simulatorc                   óˆ   — e Zd Zd„ Zd„ Zed„ ¦   «         Zed„ ¦   «         Zed„ ¦   «         Zed„ ¦   «         Z	ed„ ¦   «         Z
dS )	ÚTestCudaStreamc                 óÔ   — d„ }t          j        ¦   «         }t          j        ¦   «         }|                     ||¦  «         |                      |                     d¦  «        ¦  «         d S )Nc                 ó.   — |                      ¦   «          d S ©N)Úset)ÚstreamÚstatusÚevents      r   Úcallbackz2TestCudaStream.test_add_callback.<locals>.callback   s   € Ø�IŠI‰KŒKˆKˆKˆKr   ç      ð?)r   r   Ú	threadingÚEventÚadd_callbackÚ
assertTrueÚwait©Úselfr!   r   Úcallback_events       r   Útest_add_callbackz TestCudaStream.test_add_callback   sg   € ð	ð 	ð 	õ ”‘”ˆÝ"œÑ*Ô*ˆØ×Ò˜H nÑ5Ô5Ð5Ø�Š˜×+Ò+¨CÑ0Ô0Ñ1Ô1Ð1Ð1Ð1r   c                 óÜ   ‡ ‡— t          j        ¦   «         Šˆˆ fd„}t          j        ¦   «         }|                     |¦  «         ‰                      ‰                     d¦  «        ¦  «         d S )Nc                 óZ   •— ‰                      |¦  «         ‰                     ¦   «          d S r   )ÚassertIsNoner   )r   r   Úargr*   r)   s      €€r   r!   zCTestCudaStream.test_add_callback_with_default_arg.<locals>.callback#   s/   ø€ Ø×Ò˜cÑ"Ô"Ð"Ø×ÒÑ Ô Ð Ð Ð r   r"   )r#   r$   r   r   r%   r&   r'   r(   s   `  @r   Ú"test_add_callback_with_default_argz1TestCudaStream.test_add_callback_with_default_arg    sv   øø€ Ý"œÑ*Ô*ˆð	!ð 	!ð 	!ð 	!ð 	!ð 	!õ ”‘”ˆØ×Ò˜HÑ%Ô%Ð%Ø�Š˜×+Ò+¨CÑ0Ô0Ñ1Ô1Ð1Ð1Ð1r   c              ƒ   ód   K  — t          j        ¦   «         }|                     ¦   «         ƒ d {V —† d S r   )r   r   Ú
async_done)r)   r   s     r   Útest_async_donezTestCudaStream.test_async_done+   s;   è è € å”‘”ˆØ×ÒÑ!Ô!Ð!Ð!Ð!Ð!Ð!Ð!Ð!Ð!Ð!r   c              ƒ   óÒ   ‡ ‡K  — dt           dt           fˆ fd„Šg d¢}ˆfd„|D ¦   «         }t          j        |Ž ƒ d {V —†}‰                      t	          j        ||¦  «        ¦  «         d S )NÚvalue_inÚreturnc              “   ón  •K  — t          j        ¦   «         }t          j        d¦  «        t          j        d¦  «        }}| |d d …<   t          j        ||¬¦  «        }|                     ||¬¦  «         |                     ¦   «         ƒ d {V —†}‰                     ||¦  «         |                     ¦   «         S )Né   ©r   )r   r   Úpinned_arrayÚ	to_deviceÚcopy_to_hostr2   ÚassertEqualÚmean)r5   r   Úh_srcÚh_dstÚd_aryÚdone_resultr)   s         €r   Úasync_cuda_fnz9TestCudaStream.test_parallel_tasks.<locals>.async_cuda_fn2   s²   øè è € Ý”[‘]”]ˆFÝÔ,¨QÑ/Ô/µÔ1BÀ1Ñ1EÔ1E�5ˆEØˆE�!�!�!‰HÝ”N 5°Ð8Ñ8Ô8ˆEØ×Ò˜u¨VÐÑ4Ô4Ð4Ø &× 1Ò 1Ñ 3Ô 3Ð3Ð3Ð3Ð3Ð3Ð3ˆKØ×Ò˜[¨&Ñ1Ô1Ð1Ø—:’:‘<”<Ðr   )é   é   é   é   c                 óJ   •— g | ]}t          j         ‰|¦  «        ¦  «        ‘Œ S © )r	   Úcreate_task)Ú.0ÚvrC   s     €r   ú
<listcomp>z6TestCudaStream.test_parallel_tasks.<locals>.<listcomp>=   s.   ø€ ÐJÐJÐJ¸1•Ô$ ] ]°1Ñ%5Ô%5Ñ6Ô6ÐJÐJÐJr   )Úfloatr	   Úgatherr&   ÚnpÚallclose)r)   Ú	values_inÚtasksÚ
values_outrC   s   `   @r   Útest_parallel_tasksz"TestCudaStream.test_parallel_tasks0   sš   øøè è € ð	 ­%ð 	 µEð 	 ð 	 ð 	 ð 	 ð 	 ð 	 ð !�L�Lˆ	ØJÐJÐJÐJÀ	ÐJÑJÔJˆÝ"œ>¨5Ð1Ð1Ð1Ð1Ð1Ð1Ð1ˆ
Ø�Š�œ I¨zÑ:Ô:Ñ;Ô;Ð;Ð;Ð;r   c              ƒ   óÆ   ‡K  — t          j        ¦   «         Šˆfd„t          d¦  «        D ¦   «         }t          j        |Ž ƒ d {V —†}|D ]}|                      |‰¦  «         Œd S )Nc                 ó8   •— g | ]}‰                      ¦   «         ‘ŒS rI   ©r2   )rK   Ú_r   s     €r   rM   z;TestCudaStream.test_multiple_async_done.<locals>.<listcomp>D   s%   ø€ Ð:Ð:Ð:¨A�F×%Ò%Ñ'Ô'Ð:Ð:Ð:r   rG   )r   r   Úranger	   rO   r=   )r)   Údone_awsÚdoneÚdr   s       @r   Útest_multiple_async_donez'TestCudaStream.test_multiple_async_doneA   s�   øè è € å”‘”ˆØ:Ð:Ð:Ð:µ°q±´Ð:Ñ:Ô:ˆÝ”^ XÐ.Ð.Ð.Ð.Ð.Ð.Ð.ˆØð 	(ð 	(ˆAØ×Ò˜Q Ñ'Ô'Ð'Ð'ð	(ð 	(r   c              ƒ   óÜ   K  — d„ t          d¦  «        D ¦   «         }d„ |D ¦   «         }t          j        |Ž ƒ d {V —†}|                      t	          |¦  «        t	          |¦  «        ¦  «         d S )Nc                 ó4   — g | ]}t          j        ¦   «         ‘ŒS rI   )r   r   )rK   rY   s     r   rM   zLTestCudaStream.test_multiple_async_done_multiple_streams.<locals>.<listcomp>K   s   € Ð3Ð3Ð3 Q•4”;‘=”=Ð3Ð3Ð3r   rG   c                 ó6   — g | ]}|                      ¦   «         ‘ŒS rI   rX   )rK   r   s     r   rM   zLTestCudaStream.test_multiple_async_done_multiple_streams.<locals>.<listcomp>L   s$   € Ð>Ð>Ð>¨F�F×%Ò%Ñ'Ô'Ð>Ð>Ð>r   )rZ   r	   rO   ÚassertSetEqualr   )r)   Ústreamsr[   r\   s       r   Ú)test_multiple_async_done_multiple_streamsz8TestCudaStream.test_multiple_async_done_multiple_streamsI   sz   è è € à3Ð3­%°©(¬(Ð3Ñ3Ô3ˆØ>Ð>°gÐ>Ñ>Ô>ˆÝ”^ XÐ.Ð.Ð.Ð.Ð.Ð.Ð.ˆð 	×Ò�C ™IœI¥s¨7¡|¤|Ñ4Ô4Ð4Ð4Ð4r   c              ƒ   óT  K  — t          j        ¦   «         }|                     ¦   «         |                     ¦   «         }}|                     ¦   «          |ƒ d {V —† |                      |                     ¦   «         ¦  «         |                      |                     ¦   «         ¦  «         d S r   )r   r   r2   Úcancelr&   Ú	cancelledr\   )r)   r   Údone1Údone2s       r   Útest_cancelled_futurez$TestCudaStream.test_cancelled_futureR   sŒ   è è € å”‘”ˆØ×(Ò(Ñ*Ô*¨F×,=Ò,=Ñ,?Ô,?ˆuˆØ�Š‰ŒˆØˆˆˆˆˆˆˆØ�Š˜ŸšÑ)Ô)Ñ*Ô*Ð*Ø�Š˜Ÿ
š
™œÑ%Ô%Ð%Ð%Ð%r   N)Ú__name__Ú
__module__Ú__qualname__r+   r0   r   r3   rU   r^   rd   rj   rI   r   r   r   r      s¾   € € € € € ð2ð 2ð 2ð	2ð 	2ð 	2ð ð"ð "ñ Ôð"ð ð<ð <ñ Ôð<ð  ð(ð (ñ Ôð(ð ð5ð 5ñ Ôð5ð ð&ð &ñ Ôð&ð &ð &r   r   c                   ó>   — e Zd Zej        ed„ ¦   «         ¦   «         ZdS )ÚTestFailingStreamc              ƒ   óè  K  — t          j        ¦   «         }|                     d¦  «        }|                     d¦  «        }t          j        ¦   «         }|                     dd|¬¦  «                             ¦   «          |                     ¦   «         }|                      t          ¦  «        5  |ƒ d {V —† d d d ¦  «         n# 1 swxY w Y   |  
                    |                     ¦   «         ¦  «         d S )Nz�
            .version 6.5
            .target sm_30
            .address_size 64
            .visible .entry failing_kernel() { trap; }
        Úfailing_kernel)rD   r9   )r   Úcurrent_contextÚcreate_module_ptxÚget_functionr   Ú	configureÚ__call__r2   ÚassertRaisesÚ	ExceptionÚassertIsNotNoneÚ	exception)r)   ÚctxÚmodulerq   r   r\   s         r   Útest_failed_streamz$TestFailingStream.test_failed_streame   s'  è è € õ Ô"Ñ$Ô$ˆØ×&Ò&ð (ñ ô ˆð  ×,Ò,Ð-=Ñ>Ô>ˆå”‘”ˆØ× Ò   t°FÐ Ñ;Ô;×DÒDÑFÔFÐFØ× Ò Ñ"Ô"ˆØ×Ò�yÑ)Ô)ð 	ð 	ØˆJˆJˆJˆJˆJˆJˆJð	ð 	ð 	ñ 	ô 	ð 	ð 	ð 	ð 	ð 	ð 	øøøð 	ð 	ð 	ð 	à×Ò˜TŸ^š^Ñ-Ô-Ñ.Ô.Ð.Ð.Ð.s   Â+	C Ã CÃCN)rk   rl   rm   r   Úskipr   r}   rI   r   r   ro   ro   \   s<   € € € € € ð „]Øð/ð /ñ Ôñ „]ð/ð /ð /r   ro   Ú__main__)r	   r   r#   ÚnumpyrP   Únumbar   Únumba.cuda.testingr   r   r   r   r   ro   rk   ÚmainrI   r   r   ú<module>r„      s&  ðØ €€€Ø Ð Ð Ð Ø Ð Ð Ð Ø Ð Ð Ð Ø Ð Ð Ð Ð Ð Ø FÐ FÐ FÐ FÐ FÐ FÐ FÐ FÐ FÐ Fð	ð 	ð 	ð €Ð?Ñ@Ô@ðC&ð C&ð C&ð C&ð C&�\ñ C&ô C&ñ AÔ@ðC&ðL €Ð?Ñ@Ô@ð/ð /ð /ð /ð /˜ñ /ô /ñ AÔ@ð/ð8 ˆzÒÐØ€H„M�O„O€O€O€Oð Ðr   