o
    Zh"                     @   s  d dl Z d dlZd dlZg dZe dZe dZd dl	m
Z
 dZddgZdd Zdd	 Ze jjd
eeeeddd ZdddZe jjd
eeeede jjdeede jdg ddd Ze jjd
eeeede jjdeeddd Ze jjd
eeeede jjdeeddd Ze jjd
eeeede jjdeeddd Ze jjd
eeeede jjdeeddd ZdS )     N)uint8Zint16Zfloat32zpyarrow.cudaz
numba.cuda)DeviceNDArrayc                 C   sF   t jd t }| }t }tj|}||f||fg| _	d S )Ni  )
nprandomseedcudaContextto_numbanb_cudaZcurrent_context
from_numbacontext_choices)moduleZctx1Znb_ctx1Znb_ctx2ctx2 r   \/var/www/html/lang_env/lib/python3.10/site-packages/pyarrow/tests/test_cuda_numba_interop.pysetup_module!   s   r   c                 C   s   | ` d S N)r   )r   r   r   r   teardown_module*   s   r   c)Zidsc                 C   st   t |  \}}|j|jjksJ |j| jjksJ tj|}|j|jks(J d}||}|j|jjks8J d S )N
   )	r   handlevaluer	   r   r   r   
new_buffercontext)r   ctxnb_ctxr   sizebufr   r   r   test_context.   s   
r   hostr   c                 C   s   t |}|dkr1| dksJ t| |j }t j||d}t jjdd| t jd|dd< ||fS |dkrSt	| d|d\}}|
| |j }|j|d|jd	 ||fS td
)z5Return a host or device buffer with random data.
    r   r   dtype   )lowhighr   r!   Ndevice)targetr!   )positionnbyteszinvalid target value)r   r!   paZallocate_bufferitemsize
frombufferr   randintr   make_random_bufferr   Zcopy_from_hostr   
ValueError)r   r&   r!   r   r   arrdbufr   r   r   r-   ;   s   
r-   r!   r   )r         i  c                 C   s  t |  \}}t|d||d\}}t|}||}|j|jks"J tj| |d}	tj	
||	 |dkrBt|d d d t|d |d  d fD ]}
|||
 }tj| |d}	tj	
||
 |	 qJ	 tjtdd	 ||d d d
  W d    n1 sw   Y  |d }|| }|| |ksJ ||||}|j|jksJ tj| |d}	tj	
||	 tjtdd	 ||||d d d d d
f  W d    n1 sw   Y  d}|d }|||  }|| | |ksJ |||||}|j|jksJ tj| |d}	tj	
||	 tjtdd	 |||||d d d
  W d    n	1 s=w   Y  G dd d}|||}|j|jksYJ tj| |d}	tj	
||	 d S )Nr%   r&   r!   r   r    r2      r   zarray data is non-contiguous)match   c                   @   s    e Zd Zdd Zedd ZdS )ztest_from_object.<locals>.MyObjc                 S   s
   || _ d S r   )darr)selfr8   r   r   r   __init__   s   
z(test_from_object.<locals>.MyObj.__init__c                 S   s   | j jS r   )r8   __cuda_array_interface__)r9   r   r   r   r;      s   z8test_from_object.<locals>.MyObj.__cuda_array_interface__N)__name__
__module____qualname__r:   propertyr;   r   r   r   r   MyObj   s    r@   )r   r-   r
   Z	to_deviceZbuffer_from_objectr   r   r+   copy_to_hosttestingassert_equalslicepytestZraisesr.   Zreshape)r   r!   r   r   r   r/   cbufr8   Zcbuf2arr2sZrdarrs1s2Zs3r@   r   r   r   test_from_objectN   sd   


& rK   c           	      C   s   t |  \}}t|}d}|||j }t|f|jf||d}d|d d< d|dd < tj| d d d tj| dd  d t	j
|}tj| |d}tj||  d S )Nr   Zgpu_datac      X   r    )r   r   r!   Zmemallocr*   r   rB   rC   rA   r   Z
CudaBufferr   r+   )	r   r!   r   r   r   memr8   rF   rG   r   r   r   test_numba_memalloc   s   
rQ   c           	      C   sX   t |  \}}d}t|d||d\}}| }t|j|j|j|d}tj	|
 | d S )Nr   r%   r3   rL   )r   r-   r	   r   shapestridesr!   r   rB   rC   rA   )	r   r!   r   r   r   r/   rF   rP   r8   r   r   r   test_pyarrow_memalloc   s   rT   c           
      C   s   t |  \}}d}tjd O t|d||d\}}|jj|jjks"J | }t|j	|j
|j|d}tj| | d|d< |j  tj| |d}	|	d dksTJ W d    d S 1 s_w   Y  d S )Nr   r   r%   r3   rL   rM   r    )r   r
   Zgpusr-   r   r   r   r	   r   rR   rS   r!   r   rB   rC   rA   synchronizer+   )
r   r!   r   r   r   r/   rF   rP   r8   rG   r   r   r   test_numba_context   s   

"rV   c                 C   s   t |  \}}tjdd }d}t|d||d\}}d}|j|d  | }	| }
t|j|j|j	|
d}||	|f | |j
  tj| |j	d	}tj||d  d S )
Nc                 S   s,   t d}|| jk r| |  d7  < d S d S )Nr1   )r
   gridr   )Zan_arrayposr   r   r   increment_by_one   s   

z*test_pyarrow_jit.<locals>.increment_by_oner   r%   r3       r1   rL   r    )r   r
   Zjitr-   r   r	   r   rR   rS   r!   r   rU   r   r+   rA   rB   rC   )r   r!   r   r   rY   r   r/   rF   ZthreadsperblockZblockspergridrP   r8   Zarr1r   r   r   test_pyarrow_jit   s   

r[   )r   r   N)rE   Zpyarrowr)   numpyr   ZdtypesZimportorskipr   r
   Znumba.cuda.cudadrv.devicearrayr   r   Zcontext_choice_idsr   r   markZparametrizerangelenr   r-   rK   rQ   rT   rV   r[   r   r   r   r   <module>   sT   

	

K