
    `i                         d dl Z d dlmZmZ d dlmZ  ed           G d de                      Zedk    r e j                     dS dS )    N)CUDATestCaseskip_on_cudasim)captured_stdoutz4cudasim doesn't support cuda import at non-top-levelc                   2     e Zd ZdZ fdZ fdZd Z xZS )TestReductionz&
    Test shared memory reduction
    c                     t                      | _        | j                                         t                                                       d S N)r   _captured_stdout	__enter__supersetUpself	__class__s    /home/jaya/work/projects/VOICE-AGENT/VIET/agent-env/lib/python3.11/site-packages/numba/cuda/tests/doc_examples/test_reduction.pyr   zTestReduction.setUp   s;     / 1 1'')))    c                     | j                             d d d            t                                                       d S r	   )r
   __exit__r   tearDownr   s    r   r   zTestReduction.tearDown   s:    &&tT4888r   c                    dd l }ddlm ddlm                     |                    d                    }t          |          j        fd            } |df         |           t          |d                    t          t          |                    d                               |j                            |d         t          |                    d                               d S )Nr   )cuda)int32i   c                    j         j        }t          |           }||k     r                    d          }j                                      }| |         ||<                                    d}|j        j        k     rN|d|z  z  dk    r||xx         |||z            z  cc<   |dz  }                                 |j        j        k     N|dk    r||         | |<   d S d S d S )N      r   )	threadIdxxlengridsharedarraysyncthreadsblockDim)	datatidsizeishrsr   r   nelems	         r   	array_sumz2TestReduction.test_ex_reduction.<locals>.array_sum&   s   ."Ct99DTzzIIaLL k''u557C   """$-/))a!e}))CCaL0FA$$&&& $-/)) !88 #CDIII+ z( 8r   r   )numpynumbar   numba.typesr   	to_devicearanger   jitprintsumtestingassert_equal)r   npar+   r   r   r*   s       @@@r   test_ex_reductionzTestReduction.test_ex_reduction   s   %%%%%%
 NN299T??++A 
	) 	) 	) 	) 	) 	) 
	)8 		!U(Aadc"))D//""### 	
!c"))D//&:&:;;;;;r   )__name__
__module____qualname____doc__r   r   r8   __classcell__)r   s   @r   r   r      sj                 
0< 0< 0< 0< 0< 0< 0<r   r   __main__)	unittestnumba.cuda.testingr   r   numba.tests.supportr   r   r9   main r   r   <module>rD      s     < < < < < < < < / / / / / / GHH@< @< @< @< @<L @< @< IH@<F zHMOOOOO r   