§
    1@Äj¸  ã                   ó„   — 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        ¦   «          dS dS )é    )ÚcudaN)Úskip_on_cudasimÚCUDATestCasec                   ó>  — 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dS )	ÚTestMultiGPUContexté   zneed more than 1 gpusc                 ó†  — t          j        d¦  «        d„ ¦   «         }d„ }d}t          j        |t          j        ¬¦  «        }t          j        |t          j        ¬¦  «        }t           j        d         5   |d|f         ||¦  «         d d d ¦  «         n# 1 swxY w Y    |||¦  «          |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 ¦  «         n# 1 swxY w Y   d d d ¦  «         n# 1 swxY w Y    |||¦  «          |||	¦  «         t          j        |t          j        ¬¦  «        }t          j        |t          j        ¬¦  «        } |d|f         ||¦  «          |||¦  «         d S )Nzvoid(float64[:], float64[:])c                 ód   — t          j        d¦  «        }||j        k     r| |         dz   ||<   d S d S ©Né   ©r   ÚgridÚsize)ÚinpÚoutÚis      úVC:\Projects\Historical\venv\Lib\site-packages\numba/cuda/tests/cudapy/test_multigpu.pyÚcopy_plus_1z>TestMultiGPUContext.test_multigpu_context.<locals>.copy_plus_1   s6   € å”	˜!‘”ˆAØ�3”8Š|ˆ|Ø˜Qœ !™��A‘��ð ˆ|ó    c                 óL   — t           j                             | dz   |¦  «         d S r   )ÚnpÚtestingÚassert_equal)r   r   s     r   Úcheckz8TestMultiGPUContext.test_multigpu_context.<locals>.check   s$   € ÝŒJ×#Ò# C¨!¡G¨SÑ1Ô1Ð1Ð1Ð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Ô	1ð	$ð 	$ñ 
2Ô	1ð	$ð
	2ð 	2ð 	2ð ˆÝŒI�a�rœzÐ*Ñ*Ô*ˆÝŒI�a�rœzÐ*Ñ*Ô*ˆåŒY�qŒ\ð 	$ð 	$ØˆK˜˜1˜Ô˜a Ñ#Ô#Ð#ð	$ð 	$ð 	$ñ 	$ô 	$ð 	$ð 	$ð 	$ð 	$ð 	$ð 	$øøøð 	$ð 	$ð 	$ð 	$ð 	ˆˆa�‰Œˆàˆ�A�q�DÔ˜!˜QÑÔÐØˆˆa�‰ŒˆåŒY�qŒ\ð 	*ð 	*Ý”˜1¥B¤JÐ/Ñ/Ô/ˆBÝ”˜1¥B¤JÐ/Ñ/Ô/ˆBØˆK˜˜1˜Ô˜b "Ñ%Ô%Ð%å”˜1”ð *ð *Ý”Y˜q­¬
Ð3Ñ3Ô3�Ý”Y˜q­¬
Ð3Ñ3Ô3�Ø!�˜A˜q˜DÔ! " bÑ)Ô)Ð)ð*ð *ð *ñ *ô *ð *ð *ð *ð *ð *ð *øøøð *ð *ð *ð *ð	*ð 	*ð 	*ñ 	*ô 	*ð 	*ð 	*ð 	*ð 	*ð 	*ð 	*øøøð 	*ð 	*ð 	*ð 	*ð 	ˆˆb�"‰ŒˆØˆˆb�"‰ŒˆåŒI�a�rœzÐ*Ñ*Ô*ˆÝŒI�a�rœzÐ*Ñ*Ô*ˆØˆ�A�q�DÔ˜!˜QÑÔÐØˆˆa�‰ŒˆˆˆsJ   Á5BÂBÂBÃA'F>ÅAF'ÆF>Æ'F+	Æ+F>Æ.F+	Æ/F>Æ>GÇGz+Simulator does not support multiple threadsc                 ón  ‡‡‡— d„ Št          j        t          j        d¦  «        ¦  «        Šd}d g|z  Šˆˆˆfd„t	          |¦  «        D ¦   «         }|D ]}|                     ¦   «          Œ|D ]}|                     ¦   «          Œ‰D ].}t          |t          ¦  «        r|‚|  	                    |¦  «         Œ/d S )Nc                 óø   — 	 | 5  |                      ¦   «         }d d d ¦  «         n# 1 swxY w Y   t          j        |t          j        d¦  «        k    ¦  «        ||<   d S # t          $ r}|||<   Y d }~d S 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ð,ð ,ð ,ñ ,ô ,ð ,ð ,ð ,ð ,ð ,ð ,øøøð ,ð ,ð ,ð ,õ !#¤ s­b¬i¸©m¬mÒ';Ñ <Ô <�˜‘��øõ	 ð "ð "ð "Ø !�˜‘������øøøøð"øøøs2   ‚A „%™A ¥)©A ¬)­A Á
A9Á)A4Á4A9r-   c           	      ób   •— g | ]+}t          j        ‰t          j        j        ‰‰|f¬ ¦  «        ‘Œ,S ))ÚtargetÚargs)Ú	threadingÚThreadr   r!   Úcurrent)Ú.0r   r2   r3   r7   s     €€€r   ú
<listcomp>z:TestMultiGPUContext.test_multithreaded.<locals>.<listcomp>C   sT   ø€ ð -ð -ð -àõ Ô#¨4µt´yÔ7HØ79¸7ÀAð7Gð Hñ Hô Hð -ð -ð -r   )
r   Ú	to_devicer   r   ÚrangeÚstartÚjoinÚ
isinstanceÚBaseExceptionÚ
assertTrue)r"   ÚnthreadsÚthreadsÚthÚrr2   r3   r7   s        @@@r   Útest_multithreadedz&TestMultiGPUContext.test_multithreaded2   sù   øøø€ ð		=ð 		=ð 		=õ Œ^�BœI b™MœMÑ*Ô*ˆàˆØ�&˜8Ñ#ˆð-ð -ð -ð -ð -ð -å! (™OœOð-ñ -ô -ˆð ð 	ð 	ˆBØ�HŠH‰JŒJˆJˆJàð 	ð 	ˆBØ�GŠG‰IŒIˆIˆIàð 	#ð 	#ˆAÝ˜!�]Ñ+Ô+ð #Ø�à—’ Ñ"Ô"Ð"Ð"ð		#ð 	#r   c                 óÖ  — t           j        d„ ¦   «         }t          j        dt          j        ¬¦  «        }t           j        d         5  t          j        |¦  «        }d d d ¦  «         n# 1 swxY w Y   t           j        d         5  t          j        |¦  «        }d d d ¦  «         n# 1 swxY w Y   t           j        d         5   |d         |d¦  «         d d d ¦  «         n# 1 swxY w Y   t           j        d         5   |d         |d¦  «         d d d ¦  «         n# 1 swxY w Y   t           j        d         5  t          j                             | 	                    ¦   «         |dz   ¦  «         d d d ¦  «         n# 1 swxY w Y   t           j        d         5  t          j                             | 	                    ¦   «         |dz   ¦  «         d d d ¦  «         d S # 1 swxY w Y   d S )Nc                 óh   — t          j        d¦  «        }|| j        k     r| |xx         |z  cc<   d S d S r   r   )r5   Úvalr   s      r   Úvector_add_scalarz@TestMultiGPUContext.test_with_context.<locals>.vector_add_scalarU   s:   € å”	˜!‘”ˆAØ�3”8Š|ˆ|Ø�A��”˜#‘��‘��ð ˆ|r   r-   r   r   r   )r   r-   r   )
r   r   r   r   Úfloat32r!   r@   r   r   r.   )r"   rO   ÚhostarrÚarr1Úarr2s        r   Útest_with_contextz%TestMultiGPUContext.test_with_contextR   sY  € õ 
Œð	ð 	ñ 
Œð	õ
 ”)˜B¥b¤jÐ1Ñ1Ô1ˆÝŒY�qŒ\ð 	+ð 	+Ý”> 'Ñ*Ô*ˆDð	+ð 	+ð 	+ñ 	+ô 	+ð 	+ð 	+ð 	+ð 	+ð 	+ð 	+øøøð 	+ð 	+ð 	+ð 	+õ ŒY�qŒ\ð 	+ð 	+Ý”> 'Ñ*Ô*ˆDð	+ð 	+ð 	+ñ 	+ô 	+ð 	+ð 	+ð 	+ð 	+ð 	+ð 	+øøøð 	+ð 	+ð 	+ð 	+õ ŒY�qŒ\ð 	.ð 	.Ø$Ð˜eÔ$ T¨1Ñ-Ô-Ð-ð	.ð 	.ð 	.ñ 	.ô 	.ð 	.ð 	.ð 	.ð 	.ð 	.ð 	.øøøð 	.ð 	.ð 	.ð 	.õ ŒY�qŒ\ð 	.ð 	.Ø$Ð˜eÔ$ T¨1Ñ-Ô-Ð-ð	.ð 	.ð 	.ñ 	.ô 	.ð 	.ð 	.ð 	.ð 	.ð 	.ð 	.øøøð 	.ð 	.ð 	.ð 	.õ ŒY�qŒ\ð 	Hð 	HÝŒJ×#Ò# D×$5Ò$5Ñ$7Ô$7¸'ÀA¹+ÑGÔGÐGð	Hð 	Hð 	Hñ 	Hô 	Hð 	Hð 	Hð 	Hð 	Hð 	Hð 	Høøøð 	Hð 	Hð 	Hð 	Hõ ŒY�qŒ\ð 	Hð 	HÝŒJ×#Ò# D×$5Ò$5Ñ$7Ô$7¸'ÀA¹+ÑGÔGÐGð	Hð 	Hð 	Hñ 	Hô 	Hð 	Hð 	Hð 	Hð 	Hð 	Hð 	Hð 	Høøøð 	Hð 	Hð 	Hð 	Hð 	Hð 	Hsl   ÁA)Á)A-Á0A-ÂB'Â'B+Â.B+ÃC#Ã#C'Ã*C'Ä DÄD#Ä&D#Ä<6E>Å>FÆFÆ6GÇG"Ç%G"c                 ó  — t           j        d         5  t          j        ¦   «         }|                     d¦  «        s|                      d¦  «         d d d ¦  «         n# 1 swxY w Y   t          j        dt
          j        ¬¦  «        }t           j        d         5  t          j        |¦  «        }d d d ¦  «         n# 1 swxY w Y   t           j        d         5  t          j        t          j	        |¦  «        ¦  «        }d d d ¦  «         n# 1 swxY w Y   t           j        d         5  | 
                    |¦  «         t
          j                             |                     ¦   «         |¦  «         d d d ¦  «         d S # 1 swxY w Y   d S )Nr   r   z!Peer access between GPUs disabledr-   r   )r   r!   Úcurrent_contextÚcan_access_peerÚskipTestr   r   rP   r@   Ú
zeros_likeÚcopy_to_devicer   r   r.   )r"   ÚctxrQ   rR   rS   s        r   Útest_with_context_peer_copyz/TestMultiGPUContext.test_with_context_peer_copyn   st  € õ
 ŒY�qŒ\ð 	Cð 	CÝÔ&Ñ(Ô(ˆCØ×&Ò& qÑ)Ô)ð CØ—’ÐAÑBÔBÐBð	Cð 	Cð 	Cñ 	Cô 	Cð 	Cð 	Cð 	Cð 	Cð 	Cð 	Cøøøð 	Cð 	Cð 	Cð 	Cõ ”)˜B¥b¤jÐ1Ñ1Ô1ˆõ ŒY�qŒ\ð 	+ð 	+Ý”> 'Ñ*Ô*ˆDð	+ð 	+ð 	+ñ 	+ô 	+ð 	+ð 	+ð 	+ð 	+ð 	+ð 	+øøøð 	+ð 	+ð 	+ð 	+õ ŒY�qŒ\ð 	:ð 	:Ý”>¥"¤-°Ñ"8Ô"8Ñ9Ô9ˆDð	:ð 	:ð 	:ñ 	:ô 	:ð 	:ð 	:ð 	:ð 	:ð 	:ð 	:øøøð 	:ð 	:ð 	:ð 	:õ ŒY�qŒ\ð 	Bð 	Bà×Ò Ñ%Ô%Ð%õ ŒJ×#Ò# D×$5Ò$5Ñ$7Ô$7¸ÑAÔAÐAð	Bð 	Bð 	Bñ 	Bô 	Bð 	Bð 	Bð 	Bð 	Bð 	Bð 	Bð 	Bøøøð 	Bð 	Bð 	Bð 	Bð 	Bð 	BsH   “>AÁA!Á$A!ÂB;Â;B?ÃB?Ã'DÄDÄDÄ(AE=Å=FÆFN)Ú__name__Ú
__module__Ú__qualname__ÚunittestÚskipIfÚlenr   r!   r*   r   rK   rT   r\   © r   r   r   r      sô   € € € € € Ø€X„_�S�S˜œ‘^”^ aÒ'Ð)@ÑAÔAð&ð &ñ BÔAð&ðP €_ÐBÑCÔCð#ð #ñ DÔCð#ð> €X„_�S�S˜œ‘^”^ aÒ'Ð)@ÑAÔAðHð Hñ BÔAðHð6 €X„_�S�S˜œ‘^”^ aÒ'Ð)@ÑAÔAðBð Bñ BÔAðBð Bð Br   r   Ú__main__)Únumbar   Únumpyr   Únumba.cuda.testingr   r   r;   r`   r   r]   Úmainrc   r   r   ú<module>ri      s±   ðØ Ð Ð Ð Ð Ð Ø Ð Ð Ð Ø <Ð <Ð <Ð <Ð <Ð <Ð <Ð <Ø Ð Ð Ð Ø €€€ð@Bð @Bð @Bð @Bð @B˜,ñ @Bô @Bð @BðF ˆzÒÐØ€H„M�O„O€O€O€Oð Ðr   