U
    ?h $                     @   s   d dl Zd dlmZ d dlmZ d dlmZ d dlmZm	Z	m
Z
mZ d dlmZmZ d dlmZ d dlmZ d dlZe	e	e	e
e
e
eeegZejej
ej	fZd	Zd
ZedG dd deZedkre  dS )    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                   @   s   e Zd ZdZdd Zdd Zdd Zdd	 Zd
d Zdd Z	dd Z
dd Zdd Zdd Zdd Zdd Zdd Zdd Zdd Zd d! Zd"d# Zd$S )%TestCUDAVectorizeiAB c                 C   s:   t tdddd }d}d}|||}| |||  d S )Nr   targetc                 S   s   | | S N abr   r   X/var/www/html/venv/lib/python3.8/site-packages/numba/cuda/tests/cudapy/test_vectorize.py
vector_add.   s    z1TestCUDAVectorize.test_scalar.<locals>.vector_addg333333?gffffff@)r   
signaturesassertEqual)selfr   r   r   cr   r   r   test_scalar,   s    


zTestCUDAVectorize.test_scalarc                 C   sl   t tdddd }tD ]N}tjtj| j|d}t||}|||}tj	|| | 
|j| qd S )Nr   r   c                 S   s   | | S r   r   r   r   r   r   r   9   s    z-TestCUDAVectorize.test_1d.<locals>.vector_adddtype)r   r   dtypesnparrayrandomNaddtestingassert_allcloser   r!   )r   r   tydataexpectedactualr   r   r   test_1d7   s    


zTestCUDAVectorize.test_1dc           	      C   s   t tdddd }t }tD ]f}tjtj| j|d}t	||}||||d}|
 }t||}tj|| | |j| q d S )Nr   r   c                 S   s   | | S r   r   r   r   r   r   r   F   s    z3TestCUDAVectorize.test_1d_async.<locals>.vector_addr    stream)r   r   r   r0   r"   r#   r$   r%   r&   	to_devicecopy_to_hostr'   r(   r)   r   r!   )	r   r   r0   r*   r+   Zdevice_dataZdresultr-   r,   r   r   r   test_1d_asyncD   s    

zTestCUDAVectorize.test_1d_asyncc           
      C   s   t tdddd }ttddttD ]b\}}}d| }tj||}tj	|j
|d}|| }|||}	tj||	 | |	j| q&d S )	Nr   r   c                 S   s   | | S r   r   r   r   r   r   r   Z   s    z-TestCUDAVectorize.test_nd.<locals>.vector_add   r   )   )order)r   r   r   ranger"   ordersr#   r%   astyper$   Tr(   r)   r   r!   )
r   r   ndr!   r6   shaper+   Zdata2r,   r-   r   r   r   test_ndX   s    


zTestCUDAVectorize.test_ndc                 C   sv   t tdddd }tjdtjd}tjdtjd}|| }t|}||||d tj|| | |j	|j	 d S )Nr   r   c                 S   s   | | S r   r   r   r   r   r   r   i   s    z5TestCUDAVectorize.test_output_arg.<locals>.vector_add
   r    out)
r   r   r#   aranger   Z
empty_liker(   r)   r   r!   )r   r   ABr,   r-   r   r   r   test_output_argh   s    


z!TestCUDAVectorize.test_output_argc                 C   sh   t tdddd }tj}tD ]D}tj||d}tj|}||}tj	|| | 
||j qd S )Nr   r   c                 S   s   | | S r   r   r   r   r   r   r   x   s    z1TestCUDAVectorize.test_reduce.<locals>.vector_addr    )r   r   r#   r   input_sizesrA   r'   reducer(   r)   r   r!   )r   r   r!   nxr,   r-   r   r   r   test_reducew   s    


zTestCUDAVectorize.test_reducec           	      C   s   t tdddd }t }tj}tD ]T}tj||d}tj	|}t
||}|j	||d}tj|| | ||j q&d S )Nr   r   c                 S   s   | | S r   r   r   r   r   r   r      s    z7TestCUDAVectorize.test_reduce_async.<locals>.vector_addr    r/   )r   r   r   r0   r#   r   rE   rA   r'   rF   r1   r(   r)   r   r!   )	r   r   r0   r!   rG   rH   r,   dxr-   r   r   r   test_reduce_async   s    

z#TestCUDAVectorize.test_reduce_asyncc                 C   sj   t tdddd }d}tj|tjd}t|}|| }||| }tj	|| | 
|j|j d S )Nr   r   c                 S   s   | | S r   r   r   r   r   r   r      s    z:TestCUDAVectorize.test_manual_transfer.<locals>.vector_addr>   r    )r   r   r#   rA   r   r   r1   r2   r(   assert_equalr   r!   r   r   rG   rH   rJ   r,   r-   r   r   r   test_manual_transfer   s    


z&TestCUDAVectorize.test_manual_transferc                 C   sz   t tdddd }d}tj|tjddd}t|}||||d	 || }| }tj	
|| | |j|j d S )
Nr   r   c                 S   s   | | S r   r   r   r   r   r   r      s    z:TestCUDAVectorize.test_ufunc_output_2d.<locals>.vector_addr>   r          r?   )r   r   r#   rA   r   Zreshaper   r1   r2   r(   rL   r   r!   rM   r   r   r   test_ufunc_output_2d   s    


z&TestCUDAVectorize.test_ufunc_output_2dc                 C   s@   t tdddd }|||}tjt|t| | d S )Nr   r   c                 S   s   | | S r   r   r   r   r   r   r      s    z5TestCUDAVectorize.check_tuple_arg.<locals>.vector_add)r   r   r#   r(   rL   Zasarray)r   r   r   r   rr   r   r   check_tuple_arg   s    


z!TestCUDAVectorize.check_tuple_argc                 C   s   d}d}|  || d S )N)      ?       @      @)      @      @      @)rS   )r   r   r   r   r   r   test_tuple_arg   s    z TestCUDAVectorize.test_tuple_argc                 C   s6   t dd}|dddd}|dddd}| || d S )	NPointrH   yzrT   rU   rV   rW   rX   rY   r   rS   r   r[   r   r   r   r   r   test_namedtuple_arg   s    
z%TestCUDAVectorize.test_namedtuple_argc                 C   s<   t jdt jd}||d f}|d |d f}| || d S )Nr>   r    r4   rO   )r#   rA   r   rS   )r   Zarrr   r   r   r   r   test_tuple_of_array_arg   s    z)TestCUDAVectorize.test_tuple_of_array_argc                 C   sR   t dd}|dddd|ddddf}|d	d
dd|ddddf}| || d S )Nr[   r\   rT   rU   rV   g      ?g      @g      @rW   rX   rY   g      @g      @g      @r_   r`   r   r   r   test_tuple_of_namedtuple_arg   s    
z.TestCUDAVectorize.test_tuple_of_namedtuple_argc                 C   sf   t jdt jd}|d }t jdt jdd }|d }tdd}|||d}|||d}| || d S )Nr>   r    rO   r4   Points)ZxsZys)r#   rA   r   r   rS   )r   Zxs1Zys1Zxs2Zys2rd   r   r   r   r   r   test_namedtuple_of_array_arg   s    
z.TestCUDAVectorize.test_namedtuple_of_array_argc                 C   s&   t ddddd }| |jd d S )Nzf8(f8)r   r   c                 S   s   | d S )NrO   r   )rH   r   r   r   bar   s    z2TestCUDAVectorize.test_name_attribute.<locals>.barrf   )r   r   __name__)r   rf   r   r   r   test_name_attribute   s    

z%TestCUDAVectorize.test_name_attributec              
   C   s   t jddddt j}t|}dd }ttdd }ttdd }t	td| t	td| | 
td |  W 5 Q R X | 
td tdg W 5 Q R X z"td	gd
ddd }|| W 5 |d k	rt	td| nt`|d k	rt	td| nt`X d S )Nr4      @   c                  _   s   t ddd S )Ni  Transfer not allowed)r	   )argskwargsr   r   r   raising_transfer   s    zLTestCUDAVectorize.test_no_transfer_for_device_data.<locals>.raising_transfercuMemcpyHtoDcuMemcpyDtoHrk   zfloat32(float32)r   r   c                 S   s   | d S )NrT   r   )noiser   r   r   func  s    z@TestCUDAVectorize.test_no_transfer_for_device_data.<locals>.func)r#   r%   Zrandnr9   r   r   r1   getattrr
   setattrassertRaisesRegexr	   r2   ro   rp   r   )r   rq   rn   Zold_HtoDZold_DtoHrr   r   r   r    test_no_transfer_for_device_data   s*    

z2TestCUDAVectorize.test_no_transfer_for_device_dataN)rg   
__module____qualname__r&   r   r.   r3   r=   rD   rI   rK   rN   rQ   rS   rZ   ra   rb   rc   re   rh   rv   r   r   r   r   r   %   s$   
r   __main__)numpyr#   collectionsr   	itertoolsr   Znumbar   r   r   r   r   Znumba.cuda.cudadrv.driverr	   r
   Znumba.cuda.testingr   r   Zunittestr   r"   r8   rE   r   rg   mainr   r   r   r   <module>   s(   	 u