
    |h $                     V   S SK rS SKJr  S SKJr  S SKJr  S SKJrJ	r	J
r
Jr  S SKJrJr  S SKJr  S SKJr  S SKr\	" \	\	5      \
" \
\
5      \" \\5      /r\R                  \R                  \R                  4rS	rS
r\" S5       " S S\5      5       r\S:X  a  \R2                  " 5         gg)    N)
namedtuple)product)	vectorize)cudaint32float32float64)CudaAPIErrordriver)skip_on_cudasim)CUDATestCase)CF)   d   i  z&ufunc API unsupported in the simulatorc                   ~    \ rS rSrSrS rS rS rS rS r	S r
S	 rS
 rS rS rS rS rS rS rS rS rS rSrg)TestCUDAVectorize%   iAB c                 l    [        [        SS9S 5       nSnSnU" X#5      nU R                  XBU-   5        g )Nr   targetc                 
    X-   $ N abs     `/home/james-whalen/.local/lib/python3.13/site-packages/numba/cuda/tests/cudapy/test_vectorize.py
vector_add1TestCUDAVectorize.test_scalar.<locals>.vector_add.   	    5L    g333333?gffffff@)r   
signaturesassertEqual)selfr   r   r   cs        r   test_scalarTestCUDAVectorize.test_scalar,   sC    	:f	-	 
.	 qE"r"   c                 p   [        [        SS9S 5       n[         H  n[        R                  " [        R
                  R                  U R                  5      US9n[        R                  " X35      nU" X35      n[        R                  R                  XE5        U R                  UR                  U5        M     g )Nr   r   c                 
    X-   $ r   r   r   s     r   r   -TestCUDAVectorize.test_1d.<locals>.vector_add9   r!   r"   dtype)r   r#   dtypesnparrayrandomNaddtestingassert_allcloser$   r-   )r%   r   tydataexpectedactuals         r   test_1dTestCUDAVectorize.test_1d7   s    	:f	-	 
.	 B88BII,,TVV4B?Dvvd)H+FJJ&&x8V\\2. r"   c                    [        [        SS9S 5       n[        R                  " 5       n[         H  n[
        R                  " [
        R                  R                  U R                  5      US9n[        R                  " XB5      nU" XUUS9nUR                  5       n[
        R                  " XD5      n[
        R                  R                  X5        U R                  UR                  U5        M     g )Nr   r   c                 
    X-   $ r   r   r   s     r   r   3TestCUDAVectorize.test_1d_async.<locals>.vector_addF   r!   r"   r,   stream)r   r#   r   r@   r.   r/   r0   r1   r2   	to_devicecopy_to_hostr3   r4   r5   r$   r-   )	r%   r   r@   r6   r7   device_datadresultr9   r8   s	            r   test_1d_asyncTestCUDAVectorize.test_1d_asyncD   s    	:f	-	 
.	 B88BII,,TVV4B?D..6K &IG))+Fvvd)HJJ&&x8V\\2. r"   c                    [        [        SS9S 5       n[        [        SS5      [        [
        5       H  u  p#nSU-  n[        R                  R                  U5      R                  U5      n[        R                  " UR                  US9nXg-   nU" Xg5      n	[        R                  R                  X5        U R                  U	R                  U5        M     g )Nr   r   c                 
    X-   $ r   r   r   s     r   r   -TestCUDAVectorize.test_nd.<locals>.vector_addZ   r!   r"      r   )   )order)r   r#   r   ranger.   ordersr/   r1   astyper0   Tr4   r5   r$   r-   )
r%   r   ndr-   rL   shaper7   data2r8   r9   s
             r   test_ndTestCUDAVectorize.test_ndX   s    	:f	-	 
.	 !(aVV DBu2IE99##E*11%8DHHTVV51E|H,FJJ&&x8V\\51 !Er"   c                    [        [        SS9S 5       n[        R                  " S[        R                  S9n[        R                  " S[        R                  S9nX#-   n[        R
                  " U5      nU" X#US9  [        R                  R                  XE5        U R                  UR                  UR                  5        g )Nr   r   c                 
    X-   $ r   r   r   s     r   r   5TestCUDAVectorize.test_output_arg.<locals>.vector_addi   r!   r"   
   r,   out)
r   r#   r/   aranger   
empty_liker4   r5   r$   r-   )r%   r   ABr8   r9   s         r   test_output_arg!TestCUDAVectorize.test_output_argh   s    	:f	-	 
.	 IIb

+IIb

+5q!1V$


""846r"   c                 b   [        [        SS9S 5       n[        R                  n[         H  n[        R
                  " X2S9n[        R                  R                  U5      nUR                  U5      n[        R                  R                  XV5        U R                  X&R                  5        M     g )Nr   r   c                 
    X-   $ r   r   r   s     r   r   1TestCUDAVectorize.test_reduce.<locals>.vector_addx   r!   r"   r,   )r   r#   r/   r   input_sizesr\   r3   reducer4   r5   r$   r-   )r%   r   r-   nxr8   r9   s          r   test_reduceTestCUDAVectorize.test_reducew   s    	:f	-	 
.	 A		!)Avv}}Q'H&&q)FJJ&&x8
 ULL1 r"   c                    [        [        SS9S 5       n[        R                  " 5       n[        R
                  n[         H  n[        R                  " XCS9n[        R                  R                  U5      n[        R                  " XR5      nUR                  XrS9n[        R                  R                  Xh5        U R                  X8R                  5        M     g )Nr   r   c                 
    X-   $ r   r   r   s     r   r   7TestCUDAVectorize.test_reduce_async.<locals>.vector_add   r!   r"   r,   r?   )r   r#   r   r@   r/   r   re   r\   r3   rf   rA   r4   r5   r$   r-   )	r%   r   r@   r-   rg   rh   r8   dxr9   s	            r   test_reduce_async#TestCUDAVectorize.test_reduce_async   s    	:f	-	 
.	 A		!)Avv}}Q'H*B&&r&9FJJ&&x8ULL1 r"   c                 `   [        [        SS9S 5       nSn[        R                  " U[        R                  S9n[
        R                  " U5      nX3-   nU" X45      R                  5       n[        R                  R                  XV5        U R                  UR                  UR                  5        g )Nr   r   c                 
    X-   $ r   r   r   s     r   r   :TestCUDAVectorize.test_manual_transfer.<locals>.vector_add   r!   r"   rY   r,   )r   r#   r/   r\   r   r   rA   rB   r4   assert_equalr$   r-   r%   r   rg   rh   rn   r8   r9   s          r   test_manual_transfer&TestCUDAVectorize.test_manual_transfer   s    	:f	-	 
.	 IIarxx(^^A5A"//1


16r"   c                    [        [        SS9S 5       nSn[        R                  " U[        R                  S9R                  SS5      n[        R                  " U5      nU" XDUS9  X3-   nUR                  5       n[        R                  R                  XV5        U R                  UR                  UR                  5        g )	Nr   r   c                 
    X-   $ r   r   r   s     r   r   :TestCUDAVectorize.test_ufunc_output_2d.<locals>.vector_add   r!   r"   rY   r,         rZ   )r   r#   r/   r\   r   reshaper   rA   rB   r4   rt   r$   r-   ru   s          r   test_ufunc_output_2d&TestCUDAVectorize.test_ufunc_output_2d   s    	:f	-	 
.	 IIarxx(00A6^^A2r"5"


16r"   c                     [        [        SS9S 5       nU" X5      n[        R                  R	                  [        R
                  " U5      [        R
                  " U5      -   U5        g )Nr   r   c                 
    X-   $ r   r   r   s     r   r   5TestCUDAVectorize.check_tuple_arg.<locals>.vector_add   r!   r"   )r   r#   r/   r4   rt   asarray)r%   r   r   r   rs        r   check_tuple_arg!TestCUDAVectorize.check_tuple_arg   sP    	:f	-	 
.	 q




1

1 =qAr"   c                 .    SnSnU R                  X5        g )N)      ?       @      @)      @      @      @)r   )r%   r   r   s      r   test_tuple_arg TestCUDAVectorize.test_tuple_arg   s    Q"r"   c                 ^    [        SS5      nU" SSSS9nU" SSSS9nU R                  X#5        g )	NPointrh   yzr   r   r   r   r   r   r   r   r%   r   r   r   s       r   test_namedtuple_arg%TestCUDAVectorize.test_namedtuple_arg   s8    7O4C3#&C3#&Q"r"   c                     [         R                  " S[         R                  S9nXS-   4nUS-   US-   4nU R                  X#5        g )NrY   r,   rJ   r{   )r/   r\   r   r   )r%   arrr   r   s       r   test_tuple_of_array_arg)TestCUDAVectorize.test_tuple_of_array_arg   sA    ii"((+'N1WcAgQ"r"   c                 ~    [        SS5      nU" SSSS9U" SSSS94nU" S	S
SS9U" SSSS94nU R                  X#5        g )Nr   r   r   r   r   g      ?g      @g      @r   r   r   g      @g      @g      @r   r   s       r   test_tuple_of_namedtuple_arg.TestCUDAVectorize.test_tuple_of_namedtuple_arg   sT    7O4SC3's)CDSC3's)CDQ"r"   c                     [         R                  " S[         R                  S9nUS-   n[         R                  " S[         R                  S9S-  nUS-   n[        SS5      nU" XS9nU" X4S9nU R	                  Xg5        g )NrY   r,   r{   rJ   Points)xsys)r/   r\   r   r   r   )r%   xs1ys1xs2ys2r   r   r   s           r   test_namedtuple_of_array_arg.TestCUDAVectorize.test_namedtuple_of_array_arg   so    ii"((+Agii"((+a/AgHl3c"c"Q"r"   c                 \    [        SSS9S 5       nU R                  UR                  S5        g )Nzf8(f8)r   r   c                     U S-  $ )Nr{   r   )rh   s    r   bar2TestCUDAVectorize.test_name_attribute.<locals>.bar   s    6Mr"   r   )r   r$   __name__)r%   r   s     r   test_name_attribute%TestCUDAVectorize.test_name_attribute   s1    	8F	+	 
,	 	u-r"   c                 `   [         R                  R                  SSSS5      R                  [         R                  5      n[
        R                  " U5      nS n[        [        SS 5      n[        [        SS 5      n[        [        SU5        [        [        SU5        U R                  [        S5         UR                  5         S S S 5        U R                  [        S5         [
        R                  " S/5        S S S 5         [        S/S	S
9S 5       nU" U5        Ub  [        [        SU5        O[        ?Ub  [        [        SU5        g [        ?g ! , (       d  f       N= f! , (       d  f       Np= f! Ub  [        [        SU5        O[        ?Ub  [        [        SU5        f [        ?f = f)NrJ      @   c                      [        SS5      e)Ni  Transfer not allowed)r
   )argskwargss     r   raising_transferLTestCUDAVectorize.test_no_transfer_for_device_data.<locals>.raising_transfer   s    s$:;;r"   cuMemcpyHtoDcuMemcpyDtoHr   zfloat32(float32)r   r   c                     U S-   $ )Nr   r   )noises    r   func@TestCUDAVectorize.test_no_transfer_for_device_data.<locals>.func  s    s{"r"   )r/   r1   randnrO   r   r   rA   getattrr   setattrassertRaisesRegexr
   rB   r   r   r   )r%   r   r   old_HtoDold_DtoHr   s         r    test_no_transfer_for_device_data2TestCUDAVectorize.test_no_transfer_for_device_data   sh    		1b"-44RZZ@u%	<
 6>486>48(89(89 ##L2HI  J ##L2HINNA3 J	( *+F;# <# K
 #9'#9'5 JI JI  #9'#9's$   3E"E$E5 
E!$
E258F-r   N)r   
__module____qualname____firstlineno__r2   r'   r:   rE   rT   r`   ri   ro   rv   r~   r   r   r   r   r   r   r   r   __static_attributes__r   r"   r   r   r   %   s_    
 	A	#//(2 72$2$77B#
####.0(r"   r   __main__)numpyr/   collectionsr   	itertoolsr   numbar   r   r   r   r	   numba.cuda.cudadrv.driverr
   r   numba.cuda.testingr   r   unittestr#   r.   rN   re   r   r   mainr   r"   r   <module>r      s     "   / / : . +  E5!gw'gw')
 **bjj"((	+ 
 $ 9:q( q( ;q(h zMMO r"   