
    |hO                         S SK r S SKrS SKJr  S SKJr  S SKJrJ	r	  S SKJ
r
   " S S\	5      r " S S	\	5      r " S
 S\	5      r\S:X  a  \R                  " 5         gg)    N)devicearray)cuda)unittestCUDATestCase)skip_on_cudasimc                      \ 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5      S 5       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!5      S" 5       r#S# r$\" S$5      S% 5       r%\" S$5      S& 5       r&\" S$5      S' 5       r'\" S$5      S( 5       r(\" S$5      S) 5       r)\" S$5      S* 5       r*\" S$5      S+ 5       r+\" S$5      S, 5       r,\" S$5      S- 5       r-S. r.\" S!5      S/ 5       r/S0r0g1)2TestCudaNDArray	   c                    [         R                  " SS9n[        R                  " U5        [        R
                  " S5      n[         R                  " U5      n[        R                  " U5        [        R                  " S5      n[         R                  " U5      nU R                  UR                  S5        [        R                  " U5        g )Nd   )shapegX9v?r   )
r   device_arrayr   verify_cuda_ndarray_interfacenpempty	to_deviceasarrayassertEqualndim)selfdaryarys      d/home/james-whalen/.local/lib/python3.13/site-packages/numba/cuda/tests/cudadrv/test_cuda_ndarray.pytest_device_array_interface+TestCudaNDArray.test_device_array_interface
   s      s+11$7hhsm~~c"11$7jj~~c"A&11$7    c                 @   [         R                  " S[         R                  S9nSUR                  l        U R                  UR                  R                  5        [        R                  " U5      nUR                  5       n[         R                  R                  X15        g )Nr   dtypeF)r   arangefloat32flags	writeableassertFalser   r   copy_to_hosttestingassert_array_equal)r   r   r   retrs       r   test_device_array_from_readonly/TestCudaNDArray.test_device_array_from_readonly   sh    ii2::.#		,,-~~c"  "


%%d0r   c                     [         R                  " SSS9nU R                  UR                  [        R                  " S5      5        g )N)r   f4)r   r   )r   r   r   r   r   )r   r   s     r   test_devicearray_dtype&TestCudaNDArray.test_devicearray_dtype!   s0      vT:RXXd^4r   c                 t    [         R                  " S[         R                  S9n[        R                  " USS9  g )Nr   r   F)copy)r   r    r!   r   r   )r   arrays     r   test_devicearray_no_copy(TestCudaNDArray.test_devicearray_no_copy%   s$    		#RZZ0u5)r   c                 "   [         R                  " S5      R                  SSS5      n[        R                  " U5      nU R                  UR                  UR                  5        U R                  UR                  SS  UR                  SS  5        g )N               )r   r    reshaper   r   r   r   r   r   r   s      r   test_devicearray_shape&TestCudaNDArray.test_devicearray_shape)   si    ii	"**1a3~~c"DJJ/12

127r   c                     [         R                  " S[         R                  S9nUR                  5       n[        R
                  " U5      nSUS S & UR                  U5        [         R                  R                  X5        g )Nr   r   r   	r   r    int32r0   r   r   r%   r&   r'   )r   r1   originalgpumems       r   test_devicearray TestCudaNDArray.test_devicearray/   sX    		#RXX.::<&aE"


%%e6r   c                 X   [         R                  " 5       nUR                  5          [         R                  " S[        R
                  US9nU R                  UR                  U5      R                  U5        U R                  UR                  U5        S S S 5        g ! , (       d  f       g = f)N)r7   r7   )r   stream)r   rF   auto_synchronizer   r   float64r   bind)r   rF   arrs      r   test_stream_bind TestCudaNDArray.test_stream_bind8   sw    $$&##jjC SXXf-44f=SZZ0 '&&s   A,B
B)c                     [         R                  " S5      n[        R                  " S5      nU R	                  [        U5      [        U5      5        g )N)r7   r7   r   r   r   r   r   lenr;   s      r   test_len_1dTestCudaNDArray.test_len_1dB   s7    hhtn  #S3t9-r   c                     [         R                  " S5      n[        R                  " S5      nU R	                  [        U5      [        U5      5        g )N)r7      rN   r;   s      r   test_len_2dTestCudaNDArray.test_len_2dG   s8    hhv  (S3t9-r   c                     [         R                  " S5      n[        R                  " S5      nU R	                  [        U5      [        U5      5        g )N)r7   rS      rN   r;   s      r   test_len_3dTestCudaNDArray.test_len_3dL   s8    hhy!  +S3t9-r   c                    Sn[         R                  " U[         R                  S9nUR                  5       n[        R
                  " U5      nUR                  US-  5      u  pVSUS S & U R                  [         R                  " US:H  5      5        UR                  X!S-  S  5        UR                  US US-   5        U R                  [         R                  " X#:H  5      5        g )Nr   r   r6   r   )
r   r    r@   r0   r   r   split
assertTrueallr%   )r   Nr1   rA   rB   leftrights          r   test_devicearray_partition*TestCudaNDArray.test_devicearray_partitionQ   s    		!288,::<&ll16*auz*+5a>*%a.)u012r   c                 0   Sn[         R                  " U[         R                  S9nUR                  5       n[        R
                  " U5      n[        R
                  " US-  US9  UR                  U5        [         R                  R                  X#S-  5        g )Nr   r   r6   )tor?   )r   r^   r1   rA   rB   s        r   test_devicearray_replace(TestCudaNDArray.test_devicearray_replacea   sj    		!288,::<&uqyV,E"


%%e\:r   zThis works in the simulatorc                 v   [         R                  " [        R                  " [        R                  " S5      5      R                  SSS5      5      nU R                  [        5       n[        R                  " U5        S S S 5        U R                  S[        WR                  5      5        g ! , (       d  f       N4= f)N   r7   r8   r9   z2transposing a non-2D DeviceNDArray isn't supported)r   r   r   r1   r    r:   assertRaisesNotImplementedError	transposer   str	exceptionr   rB   es      r   #test_devicearray_transpose_wrongdim3TestCudaNDArray.test_devicearray_transpose_wrongdimj   sz    2 7 ? ?1a HI23qLL  4 	@	 43s   %B**
B8c                 6   [         R                  " [         R                  " S5      5      R                  SSS5      n[         R                  " [
        R                  " U5      SS9R                  5       nU R                  [         R                  " X!:H  5      5        g )Nr5   r7   r8   r6   )r   r9   r6   axes)
r   r1   r    r:   rk   r   r   r%   r\   r]   r   rA   r1   s      r   #test_devicearray_transpose_identity3TestCudaNDArray.test_devicearray_transpose_identityu   se    88BIIbM*221a;T^^H5"+--9\^ 	u012r   c                 r   [         R                  " [        R                  " [        R                  " S5      5      R                  SS5      5      nU R                  [        5       n[        R                  " USS9  S S S 5        U R                  [        WR                  5      SS/S9  g ! , (       d  f       N4= f)	Nrh   r7   r8   )r   r   rs   zinvalid axes list (0, 0)zrepeated axis in transpose	containerr   r   r   r1   r    r:   ri   
ValueErrorrk   assertInrl   rm   rn   s      r   )test_devicearray_transpose_duplicatedaxis9TestCudaNDArray.test_devicearray_transpose_duplicatedaxis|   s    2 7 ? ?1 EFz*aLLf- + 	*, 	 	 +*   $B((
B6c                 r   [         R                  " [        R                  " [        R                  " S5      5      R                  SS5      5      nU R                  [        5       n[        R                  " USS9  S S S 5        U R                  [        WR                  5      / SQS9  g ! , (       d  f       N4= f)Nrh   r7   r8   )r   r6   rs   )zinvalid axes list (0, 2)zinvalid axis for this arrayz0axis 2 is out of bounds for array of dimension 2ry   r{   rn   s      r   $test_devicearray_transpose_wrongaxis4TestCudaNDArray.test_devicearray_transpose_wrongaxis   s~    2 7 ? ?1 EFz*aLLf- + 	 	 	 +*r   c                    [         R                  " [         R                  " S5      SS9R                  SS5      n[        R
                  " U5      nS Hg  nU R                  US9   [         R                  R                  UR                  U5      R                  5       UR                  U5      5        S S S 5        Mi     g ! , (       d  f       M{  = f)Nrh   i2r   r7   r8   )i4u4i8f8)r   r1   r    r:   r   r   subTestr&   r'   viewr%   )r   rA   r1   r   s       r   test_devicearray_view_ok(TestCudaNDArray.test_devicearray_view_ok   s    88BIIbM6>>q!Dx(-EE*

--JJu%224MM%( +* .**s   $AB>>
C	c                 f   [         R                  " [         R                  " S5      SS9R                  SS5      n[        R
                  " U5      S S 2S S S24   nUS S 2S S S24   n[         R                  R                  UR                  S5      R                  5       UR                  S5      5        g )N    r   r   r8      r6   u2)
r   r1   r    r:   r   r   r&   r'   r   r%   ru   s      r   %test_devicearray_view_ok_not_c_contig5TestCudaNDArray.test_devicearray_view_ok_not_c_contig   s    88BIIbM6>>q!Dx(CaC0AssF#


%%JJt))+MM$	
r   c                    [         R                  " [         R                  " S5      SS9R                  SS5      n[        R
                  " U5      S S 2S S S24   nU R                  [        5       nUR                  S5        S S S 5        [        WR                  5      nU R                  SU5        S	U;   nS
U;   nU R                  U=(       d    US5        g ! , (       d  f       N[= f)Nr   r   r   r8   r   r6   r   z)To change to a dtype of a different size,zthe array must be C-contiguousz the last axis must be contiguousz&Expected message to mention contiguity)r   r1   r    r:   r   r   ri   r|   r   rl   rm   r}   r\   )r   rA   r1   ro   msgcontiguous_pre_np123contiguous_post_np123s          r   &test_devicearray_view_bad_not_c_contig6TestCudaNDArray.test_devicearray_view_bad_not_c_contig   s    88BIIbM6>>q!Dx(CaC0z*aJJt + !++A3G?3F Bc I,E0E@	B +*s   /C
C$c                 l   [         R                  " [         R                  " S5      SS9R                  SS5      n[        R
                  " U5      nU R                  [        5       nUR                  S5        S S S 5        U R                  S[        WR                  5      5        g ! , (       d  f       N4= f)Nrh   r   r   r8   r7   r   zuWhen changing to a larger dtype, its size must be a divisor of the total size in bytes of the last axis of the array.)r   r1   r    r:   r   r   ri   r|   r   r   rl   rm   )r   rA   r1   ro   s       r   "test_devicearray_view_bad_itemsize2TestCudaNDArray.test_devicearray_view_bad_itemsize   s    88BIIbM6>>q!Dx(z*aJJt +. 		 +*s   %B%%
B3c                 J   [         R                  " [         R                  " S5      5      R                  SS5      n[         R                  " [
        R                  " U5      5      R                  5       nU R                  [         R                  " X!R                  :H  5      5        g Nrh   r7   r8   )r   r1   r    r:   rk   r   r   r%   r\   r]   Tru   s      r   test_devicearray_transpose_ok-TestCudaNDArray.test_devicearray_transpose_ok   sa    88BIIbM*221a8T^^H56CCEu

234r   c                 6   [         R                  " [         R                  " S5      5      R                  SS5      n[        R
                  " U5      R                  R                  5       nU R                  [         R                  " X!R                  :H  5      5        g r   )
r   r1   r    r:   r   r   r   r%   r\   r]   ru   s      r   test_devicearray_transpose_T,TestCudaNDArray.test_devicearray_transpose_T   s^    88BIIbM*221a8x(**779u

234r   c                    [         R                  " S5      R                  SSSS9n[         R                  " SSS9n[        R
                  " U5      nX!S'   U R                  [        5       nUS   R                  U5        S S S 5        U R                  [        R                  [        WR                  5      5        g ! , (       d  f       NB= f)N   rS   Forder)rS   )
fill_valuer   r6   )r   r    r:   fullr   r   ri   r|   copy_to_devicer   r   errmsg_contiguous_bufferrl   rm   )r   asdro   s        r   !test_devicearray_contiguous_slice1TestCudaNDArray.test_devicearray_contiguous_slice   s     IIbM!!!Qc!2GGq-NN1! z*aaD" +00	 +*s   *B;;
C	c                 f   U R                  UR                  R                  5        U R                  UR                  R                  5        X"4X!4X4X44 H  u  p4UR                  R                  (       a  SOS< SUR                  R                  (       a  SOS< 3n[        R
                  " U5      nUR                  U5        U R                  [        R                  " UR                  5       U:H  5      US9  U R                  [        R                  " UR                  5       U:H  5      US9  M     g)z
Checks host->device memcpys
Cr   z => )r   N)
r\   r"   c_contiguousf_contiguousr   r   r   r   r]   r%   )r   a_ca_frA   r0   r   r   s          r   &_test_devicearray_contiguous_host_copy6TestCudaNDArray._test_devicearray_contiguous_host_copy   s     			../		../ JJJJ	
NH  ~~22;zz..C7C
 x(AT"OOBFF1>>#3s#:;OEOOBFF1>>#3s#:;OE
r   c                     [         R                  " S5      R                  SSS5      n[         R                  " USS9nU R	                  X5        g )N}   rS   r   r   )r   r    r:   r1   r   r   r   r   s      r   (test_devicearray_contiguous_copy_host_3d8TestCudaNDArray.test_devicearray_contiguous_copy_host_3d   s=    ii	"**1a3hhs#&33C=r   c                 |    [         R                  " S5      n[         R                  " USS9nU R                  X5        g )NrS   r   r   )r   r    r1   r   r   s      r   (test_devicearray_contiguous_copy_host_1d8TestCudaNDArray.test_devicearray_contiguous_copy_host_1d   s-    iilhhs#&33C=r   c                 J   [         R                  " S5      R                  SSS5      n[         R                  " USS9nU R	                  UR
                  R                  5        U R	                  UR
                  R                  5        [        R                  " U5      nU R                  [        5       nUR                  [        R                  " U5      5        S S S 5        U R                  SR                  UR                  UR                  5      [!        WR"                  5      5        UR                  [        R                  " U5      5        U R	                  [         R$                  " UR'                  5       U:H  5      5        [        R                  " U5      nU R                  [        5       nUR                  [        R                  " U5      5        S S S 5        U R                  SR                  UR                  UR                  5      [!        UR"                  5      5        UR                  [        R                  " U5      5        U R	                  [         R$                  " UR'                  5       U:H  5      5        g ! , (       d  f       GN= f! , (       d  f       N= f)Nr   rS   r   r   zincompatible strides: {} vs. {})r   r    r:   r1   r\   r"   r   r   r   r   ri   r|   r   r   formatstridesrl   rm   r]   r%   )r   r   r   r   ro   s        r   'test_devicearray_contiguous_copy_device7TestCudaNDArray.test_devicearray_contiguous_copy_device  s   ii	"**1a3hhs#&		../		../NN3z*aT^^C01 +-44S[[#++N	 	
,-q~~/3678NN3z*aT^^C01 +-44S[[#++N	 	
,-q~~/3678% +* +*s   2&J/&J
J
J"c                    SnSn[         R                  " U5      n[         R                  " U5      R                  USS9n[         R                  " U5      R                  USS9n[	        [        U5      5       GHk  n[        S 5      4U-  [         R                  4-   nUS U U4-   X&S  -   n[         R                  " XG   U5      n	[         R                  " XW   U5      n
[        R                  " U	5      n[        R                  " U
5      n[         R                  R                  UR                  5       U	5        [         R                  R                  UR                  5       U
5        UR                  U
5        UR                  U	5        [         R                  R                  UR                  5       U
5        [         R                  R                  UR                  5       U	5        GMn     g )Nr8   )r6   r7   r   r   r   )r   prodr    r:   rangerO   slicenewaxisbroadcast_tor   r   r&   r'   r%   r   )r   	broadsize	coreshapecoresizecore_ccore_fdimnewindex
broadshapebroad_cbroad_fdbroad_cdbroad_fs                r   $test_devicearray_broadcast_host_copy4TestCudaNDArray.test_devicearray_broadcast_host_copy  sq   		779%8$,,Yc,B8$,,Yc,BY(Cd~+rzzm;H"4CI<7)D/IJoof&6
CGoof&6
CG~~g.H~~g.HJJ))(*?*?*A7KJJ))(*?*?*A7K##G,##G,JJ))(*?*?*A7KJJ))(*?*?*A7K )r   c                    [         R                  " S5      n[        R                  " U5      n[         R                  " S5      S S S2   nUR	                  U5        [         R
                  R                  UR                  5       U5        g )N
      r6   )r   r    r   r   r   r&   r'   r%   )r   r   r   rJ   s       r   (test_devicearray_contiguous_host_strided8TestCudaNDArray.test_devicearray_contiguous_host_strided3  s\    iimNN3iimCaC 	


%%ann&6<r   c                    [         R                  " [        R                  " S5      5      n[        R                  " S5      nU R	                  [
        5       nUR                  [         R                  " U5      S S S2   5        S S S 5        U R                  [        R                  [        WR                  5      5        g ! , (       d  f       NB= f)Nr   r6   )r   r   r   r    ri   r|   r   r   r   r   rl   rm   )r   r   rJ   ro   s       r   *test_devicearray_contiguous_device_strided:TestCudaNDArray.test_devicearray_contiguous_device_strided:  s    NN299R=)iimz*aT^^C0156 +00	 +*s   ,B>>
Cz,DeviceNDArray class not present in simulatorc                     [         R                  " SS[        R                  5      nU R	                  UR
                  S   5        U R	                  UR
                  S   5        g )N)r9   r   )i   r   C_CONTIGUOUSF_CONTIGUOUS)r   DeviceNDArrayr   rH   r\   r"   )r   rJ   s     r    test_devicearray_relaxed_strides0TestCudaNDArray.test_devicearray_relaxed_stridesD  sH     ''2::F 			.12		.12r   c                 N   SnSn[         R                  " X5       H  u  p4[        R                  " X4S9n[        R
                  " U5      nU R                  UR                  S   UR                  S   5        U R                  UR                  S   UR                  S   5        M     g )N))r9   r8   )r8   r9   )r   r   r   r   r   )	itertoolsproductr   ndarrayr   r   r   r"   )r   shapesordersr   r   rJ   d_arrs          r   !test_c_f_contiguity_matches_numpy1TestCudaNDArray.test_c_f_contiguity_matches_numpyS  s     "%--f=LE**U0CNN3'ESYY~6"[[8:SYY~6"[[8: >r   z Typing not done in the simulatorc                     [         R                  " SSS9n[        R                  " U5      nU R	                  UR
                  R                  S5        g Nr   r   r   r   zerosr   r   r   _numba_type_layoutr   r   r   s      r   &test_devicearray_typing_order_simple_c6TestCudaNDArray.test_devicearray_typing_order_simple_ca  <     HHRs#NN1..4r   c                     [         R                  " SSS9n[        R                  " U5      nU R	                  UR
                  R                  S5        g )Nr   r   r   r   r   r   s      r   &test_devicearray_typing_order_simple_f6TestCudaNDArray.test_devicearray_typing_order_simple_fh  r   r   c                     [         R                  " SSS9n[        R                  " U5      nU R	                  UR
                  R                  S5        g )Nr6   r   r   r   r   r   s      r   "test_devicearray_typing_order_2d_c2TestCudaNDArray.test_devicearray_typing_order_2d_co  <     HHWC(NN1..4r   c                     [         R                  " SSS9n[        R                  " U5      nU R	                  UR
                  R                  S5        g )Nr  r   r   r   r   s      r   "test_devicearray_typing_order_2d_f2TestCudaNDArray.test_devicearray_typing_order_2d_fv  r  r   c                     [         R                  " SSS9n[        R                  " U5      S S 2S4   nU R	                  UR
                  R                  S5        g )NrS   rS   r   r   r6   Ar   r   s      r   /test_devicearray_typing_order_noncontig_slice_c?TestCudaNDArray.test_devicearray_typing_order_noncontig_slice_c}  sE     HHV3'NN1ac"..4r   c                     [         R                  " SSS9n[        R                  " U5      SS S 24   nU R	                  UR
                  R                  S5        g )Nr	  r   r   r6   r
  r   r   s      r   /test_devicearray_typing_order_noncontig_slice_f?TestCudaNDArray.test_devicearray_typing_order_noncontig_slice_f  E     HHV3'NN1ac"..4r   c                     [         R                  " SSS9n[        R                  " U5      SS S 24   nU R	                  UR
                  R                  S5        g )Nr	  r   r   r6   r   r   s      r   ,test_devicearray_typing_order_contig_slice_c<TestCudaNDArray.test_devicearray_typing_order_contig_slice_c  r  r   c                     [         R                  " SSS9n[        R                  " U5      S S 2S4   nU R	                  UR
                  R                  S5        g )Nr	  r   r   r6   r   r   r   s      r   ,test_devicearray_typing_order_contig_slice_f<TestCudaNDArray.test_devicearray_typing_order_contig_slice_f  sE     HHV3'NN1ac"..4r   c                     [         R                  " [         R                  " S/5      S5      n[        R                  " U5      nU R                  UR                  R                  S5        g )Nr9   )r   r
  )r   r   r1   r   r   r   r   r   r   s      r   )test_devicearray_typing_order_broadcasted9TestCudaNDArray.test_devicearray_typing_order_broadcasted  sF     OOBHHaSM51NN1..4r   c                     [         R                  " S[         R                  S9n[        R                  " U5      n[         R
                  " U5      nU R                  UR                  UR                  5        g )Nr   r   )r   r    int16r   r   r   r   r   )r   r   r   gots       r   test_bug6697TestCudaNDArray.test_bug6697  sJ    ii"((+~~c"jjDJJ/r   c                    [         R                  " SS[        R                  S9n[        R                  " SS[        R                  S9nU R                  UR                  S5        UR                  U5        UR                  U5        [        R                  " U5      nU R                  UR                  S5        U R                  UR                  S5        UR                  U5        UR                  U5        g )N)r   )r   )r   r   r   )r   r   r   int8r   r   r   r%   r   r   r   r   )r   	dev_array
host_arraydev_array_from_hosts       r   test_issue_8477TestCudaNDArray.test_issue_8477  s      --D$46GG=	 ZZdDH
 	++T2 	z*  ,
 #nnZ8,22D9,44d;  !45**95r    N)1__name__
__module____qualname____firstlineno__r   r)   r-   r2   r<   rC   rK   rP   rT   rX   ra   re   r   rp   rv   r~   r   r   r   r   r   r   r   r   r   r   r   r   r   r   r   r   r   r   r   r  r  r  r  r  r  r  r  r$  __static_attributes__r&  r   r   r	   r	   	   s   815*871.
.
.
3 ; 23 43
B	5
5
(F.>
>
98L*= CD3 E3: 785 95 785 95 785 95 785 95 785 95 785 95 785 95 785 95 785 950 CD 6 E 6r   r	   c                       \ rS rSrS rSrg)TestRecarrayi  c                    [         R                  " SS[         R                  4S[         R                  4/S9n[         R                  " UR
                  [         R                  S9Ul        [         R                  " UR
                  [         R                  S9S-  Ul        UR                  nUR                  nS n[         R                  " U5      n[         R                  " U5      n[        R                  " U5      SUR
                  4   " XU5        [         R                  R                  X%5        [         R                  R                  X65        g )N)   value1value2r   r   c                     [         R                  " S5      nX0R                  :  a#  U R                  U   X'   U R                  U   X#'   g g )Nr9   )r   gridsizer0  r1  )xout1out2is       r   test(TestRecarray.test_recarray.<locals>.test  s;    		!A66z((1+((1+ r   r9   )r   recarrayint64rH   r    r4  r0  r1  
zeros_liker   jitr&   r'   )r   r   expect1expect2r9  got1got2s          r   test_recarrayTestRecarray.test_recarray  s    KKrxx rzz"&
  99QVV288499QVV2::6<((((	& }}W%}}W%q!&&y!!40


%%g4


%%g4r   r&  N)r'  r(  r)  r*  rC  r+  r&  r   r   r-  r-    s    5r   r-  c                   b    \ 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)TestCoreContiguousi  c                     U R                  [        R                  " U5      [        R                  " U5      R                  S   5        g )Nr   )r   r   is_contiguous
array_corer"   )r   r   s     r   _test_against_array_core+TestCoreContiguous._test_against_array_core  s7    %%d+""4(..~>	
r   c                 P    [         R                  " SSS9nU R                  U5        g r   r   r   rJ  r   d_as     r   test_device_array_like_1d,TestCoreContiguous.test_device_array_like_1d  "    #.%%c*r   c                 P    [         R                  " SSS9nU R                  U5        g Nr   rh   r   r   rM  rN  s     r   test_device_array_like_2d,TestCoreContiguous.test_device_array_like_2d  "    4%%c*r   c                 d    [         R                  " SSS9nU R                  UR                  5        g rT  r   r   rJ  r   rN  s     r   #test_device_array_like_2d_transpose6TestCoreContiguous.test_device_array_like_2d_transpose  &    4%%cee,r   c                 P    [         R                  " SSS9nU R                  U5        g )Nr   rh      r   r   rM  rN  s     r   test_device_array_like_3d,TestCoreContiguous.test_device_array_like_3d  "    C8%%c*r   c                 P    [         R                  " SSS9nU R                  U5        g )Nr   r   r   rM  rN  s     r   test_device_array_like_1d_f.TestCoreContiguous.test_device_array_like_1d_f  rR  r   c                 P    [         R                  " SSS9nU R                  U5        g NrU  r   r   rM  rN  s     r   test_device_array_like_2d_f.TestCoreContiguous.test_device_array_like_2d_f  rX  r   c                 d    [         R                  " SSS9nU R                  UR                  5        g rh  rZ  rN  s     r   %test_device_array_like_2d_f_transpose8TestCoreContiguous.test_device_array_like_2d_f_transpose  r]  r   c                 P    [         R                  " SSS9nU R                  U5        g )Nr_  r   r   rM  rN  s     r   test_device_array_like_3d_f.TestCoreContiguous.test_device_array_like_3d_f	  rc  r   c                 b    Sn[         R                  " U5      S S S2   nU R                  U5        g )Nr   r6   r   r   rJ  r   r   r   s      r   test_1d_viewTestCoreContiguous.test_1d_view  s+    xxss#%%d+r   c                 `    Sn[         R                  " USS9S S S2   nU R                  U5        g )Nr   r   r   r6   rr  rs  s      r   test_1d_view_f!TestCoreContiguous.test_1d_view_f  s.    xxS)#A#.%%d+r   c                 l    Sn[         R                  " U5      S S S2S S S24   nU R                  U5        g )NrU  r6   rr  rs  s      r   test_2d_viewTestCoreContiguous.test_2d_view  s3    xxssCaCx(%%d+r   c                 j    Sn[         R                  " USS9S S S2S S S24   nU R                  U5        g )NrU  r   r   r6   rr  rs  s      r   test_2d_view_f!TestCoreContiguous.test_2d_view_f  s6    xxS)#A#ss(3%%d+r   r&  N)r'  r(  r)  r*  rJ  rP  rV  r[  ra  re  ri  rl  ro  rt  rw  rz  r}  r+  r&  r   r   rF  rF    sC    
++-+++-+,
,
,
,r   rF  __main__)r   numpyr   numba.cuda.cudadrvr   numbar   numba.cuda.testingr   r   r   r	   r-  rF  r'  mainr&  r   r   <module>r     sW      *  5 .6l 6D5< 569, 9,x zMMO r   