
    gZ"                        d dl Z d dlZ	 d dlZg dZ e j                  d      Z
 e j                  d      Zd dlmZ dZddgZd Zd Ze j                  j%                  d e ee            e	      d
        ZddZe j                  j%                  d e ee            e	      e j                  j%                  dee	      e j                  j%                  dg d      d                      Ze j                  j%                  d e ee            e	      e j                  j%                  dee	      d               Ze j                  j%                  d e ee            e	      e j                  j%                  dee	      d               Ze j                  j%                  d e ee            e	      e j                  j%                  dee	      d               Ze j                  j%                  d e ee            e	      e j                  j%                  dee	      d               Zy# e$ r e j                  j                  ZY Cw xY w)    N)uint8int16float32zpyarrow.cudaz
numba.cuda)DeviceNDArrayc                 
   t         j                  j                  d       t        j	                         }|j                         }t        j                         }t        j                  j                  |      }||f||fg| _	        y )Ni  )
nprandomseedcudaContextto_numbanb_cudacurrent_context
from_numbacontext_choices)modulectx1nb_ctx1nb_ctx2ctx2s        Z/var/www/openai/venv/lib/python3.12/site-packages/pyarrow/tests/test_cuda_numba_interop.pysetup_moduler   $   s_    IINN4<<>DmmoG%%'G<<""7+D#Wog?F    c                     | ` y N)r   )r   s    r   teardown_moduler   -   s    r   c)idsc                    t         |    \  }}|j                  |j                  j                  k(  sJ |j                  |j                         j                  j                  k(  sJ t        j
                  j                  |      }|j                  |j                  k(  sJ d}|j                  |      }|j                  |j                  j                  k(  sJ y )N
   )	r   handlevaluer   r   r   r   
new_buffercontext)r   ctxnb_ctxr   sizebufs         r   test_contextr)   1   s     "!$KC::,,,,,::..44444<<""6*D::$$$D
..
C::+++++r   c                    t        j                  |      }|dk(  rx| dk\  sJ t        j                  | |j                  z        }t        j
                  ||      }t         j                  j                  dd| t         j                        |dd ||fS |dk(  rQt        | d|      \  }}|j                  | |j                  z        }|j                  |d|j                  	       ||fS t        d
      )z5Return a host or device buffer with random data.
    hostr   dtype   )lowhighr'   r-   Ndevice)targetr-   )positionnbyteszinvalid target value)r   r-   paallocate_bufferitemsize
frombufferr	   randintr   make_random_bufferr#   copy_from_hostr'   
ValueError)r'   r2   r-   r%   r(   arrdbufs          r   r:   r:   >   s     HHUOEqyy  enn!45mmCu-""qs)+ # 3ACx	8	%d6GS~~dU^^34C!CHH=Dy
+
,,r   r-   r'   )r         i  c                    t         |    \  }}t        |d||      \  }}t        j                  |      }|j	                  |      }|j
                  |j
                  k(  sJ t        j                  |j                         |      }	t        j                  j                  ||	       |dk\  rt        |dz  d d       t        |dz  |dz   d       fD ]^  }
|j	                  ||
         }t        j                  |j                         |      }	t        j                  j                  ||
   |	       ` 	 t        j                  t        d      5  |j	                  |d d d          d d d        |dz  }||z  }||z  |k(  sJ |j	                  |j                  ||            }|j
                  |j
                  k(  sJ t        j                  |j                         |      }	t        j                  j                  ||	       t        j                  t        d      5  |j	                  |j                  ||      d d d d df          d d d        d}|dz  }|||z  z  }||z  |z  |k(  sJ |j	                  |j                  |||            }|j
                  |j
                  k(  sJ t        j                  |j                         |      }	t        j                  j                  ||	       t        j                  t        d      5  |j	                  |j                  |||      d d d          d d d         G d	 d
      }|j	                   ||            }|j
                  |j
                  k(  sJ t        j                  |j                         |      }	t        j                  j                  ||	       y # 1 sw Y   ZxY w# 1 sw Y   xY w# 1 sw Y   xY w)Nr1   r2   r-   r%   r,   r@      zarray data is non-contiguous)match   c                   "    e Zd Zd Zed        Zy)test_from_object.<locals>.MyObjc                     || _         y r   )darr)selfrI   s     r   __init__z(test_from_object.<locals>.MyObj.__init__   s	    DIr   c                 .    | j                   j                  S r   )rI   __cuda_array_interface__)rJ   s    r   rM   z8test_from_object.<locals>.MyObj.__cuda_array_interface__   s    99555r   N)__name__
__module____qualname__rK   propertyrM    r   r   MyObjrG      s    	 
	6 
	6r   rS   )r   r:   r   	to_devicebuffer_from_objectr'   r   r8   copy_to_hosttestingassert_equalslicepytestraisesr<   reshape)r   r-   r'   r%   r&   r=   cbufrI   cbuf2arr2srdarrs1s2s3rS   s                   r   test_from_objectre   Q   sZ   
 "!$KC"43OIC S!D""4(E::"""==++-U;DJJC& qyat,a47T24A**473E==!3!3!5UCDJJ##CFD1	4  ]]:"@C""4!9-C
 1W2XBw$&&t||B';<zzTYY&&&}}U//1?


T*]]:"@C""4<<B#73Q3#?@C
 1WBrE]Bw|t###&&t||BB'?@zzTYY&&&}}U//1?


T*]]:"@C""4<<B#;CaC#@AC6 6 ""5;/E::"""==++-U;DJJC&YC CC CC Cs$   N<,O		)O<O	OOc                 x   t         |    \  }}t        j                  |      }d}|j                  ||j                  z        }t        |f|j                  f||      }d|d d d|dd  t        j                  j                  |j                         d d d       t        j                  j                  |j                         dd  d       t        j                  j                  |      }t        j                  |j                         |      }t        j                  j                  ||j                                y )Nr    gpu_datac      X   r,   )r   r   r-   memallocr7   r   rW   rX   rV   r   
CudaBufferr   r8   )	r   r-   r%   r&   r'   memrI   r]   r_   s	            r   test_numba_memallocro      s     "!$KCHHUOE D
//$/
0C$5>>"3USIDD!HDHJJD--/3R8JJD--/3R8 ??%%c*D==**,E:DJJD$"3"3"56r   c                    t         |    \  }}d}t        |d||      \  }}|j                         }t        |j                  |j
                  |j                  |      }t        j                  j                  |j                         |       y )Nr    r1   rB   rg   )r   r:   r   r   shapestridesr-   r   rW   rX   rV   )	r   r-   r%   r&   r'   r=   r]   rn   rI   s	            r   test_pyarrow_memallocrs      sq     "!$KCD"43OIC --/CCKKSIDJJD--/5r   c                 f   t         |    \  }}d}t        j                  d   5  t        |d||      \  }}|j                  j
                  |j
                  j                  k(  sJ |j                         }t        |j                  |j                  |j                  |      }t        j                  j                  |j                         |       d|d<   |j                  j!                          t        j"                  |j                         |      }	|	d   dk(  sJ 	 d d d        y # 1 sw Y   y xY w)Nr    r   r1   rB   rg   ri   r,   )r   r   gpusr:   r$   r!   r"   r   r   rq   rr   r-   r   rW   rX   rV   synchronizer8   )
r   r-   r%   r&   r'   r=   r]   rn   rI   r_   s
             r   test_numba_contextrw      s     "!$KCD	a&tH-2=	T||""fmm&9&9999mmoSYYSYYM


 1 1 3S9Q  "}}T..0>Aw"}} 
s   C;D''D0c                 
   t         |    \  }}t        j                  d        }d}t        |d||      \  }}d}|j                  |dz
  z   |z  }	|j                         }
t        |j                  |j                  |j                  |
      } ||	|f   |       |j                  j                          t        j                  |j                         |j                        }t        j                  j!                  ||dz          y )	Nc                 h    t         j                  d      }|| j                  k  r| |xx   dz  cc<   y y )Nr?   )r   gridr'   )an_arrayposs     r   increment_by_onez*test_pyarrow_jit.<locals>.increment_by_one   s.    ll1oSMQM r   r    r1   rB       r?   rg   r,   )r   r   jitr:   r'   r   r   rq   rr   r-   r$   rv   r   r8   rV   rW   rX   )r   r-   r%   r&   r}   r'   r=   r]   threadsperblockblockspergridrn   rI   arr1s                r   test_pyarrow_jitr      s     "!$KC[[  D"43OICOXX1!45/IM
--/CCKKSID4]O34T:LL==**,CII>DJJD#'*r   )r+   r   N)rZ   pyarrowr5   numpyr   ImportErrormark
pytestmarkdtypesimportorskipr   r   numba.cuda.cudadrv.devicearrayr   r   context_choice_idsr   r   parametrizerangelenr)   r:   re   ro   rs   rw   r   rR   r   r   <module>r      s  $  # 
'v>*
&

l
+ 8 $l3 @ eC(:$;</  1,1,-& eC(:$;</  1&f51H' 2 61H'V eC(:$;</  1&f57 617( eC(:$;</  1&f56 616 eC(:$;</  1&f5 61  eC(:$;</  1&f5+ 61+K  #""J#s   H3 3II