
    xKgc                         d dl Z d dlZd dlZd dlZd dlmZ d dlmZm	Z	m
Z
 d Z e
d       G d de	             Z e
d       G d d	e	             Zed
k(  r ej                          yy)    N)cuda)unittestCUDATestCaseskip_on_cudasimc                 B     t        j                          fd       }|S )Nc                      t        j                         }|j                  d       	 |j                   | i |      |j	                          S # |j	                          w xY w)NT)asyncionew_event_loop	set_debugrun_until_completeclose)argskwdsloopfs      i/home/alanp/www/video.onchill/myenv/lib/python3.12/site-packages/numba/cuda/tests/cudadrv/test_streams.pyrunnerz!with_asyncio_loop.<locals>.runner
   sM    %%'t	**1d+;d+;<JJLDJJLs   A A!)	functoolswraps)r   r   s   ` r   with_asyncio_loopr   	   s%    __Q  M    z,CUDA Driver API unsupported in the simulatorc                   h    e Zd Zd Zd Zed        Zed        Zed        Zed        Z	ed        Z
y)	TestCudaStreamc                     d }t        j                         }t        j                         }|j	                  ||       | j                  |j                  d             y )Nc                 $    |j                          y N)set)streamstatusevents      r   callbackz2TestCudaStream.test_add_callback.<locals>.callback   s    IIKr         ?)r   r   	threadingEventadd_callback
assertTruewaitselfr!   r   callback_events       r   test_add_callbackz TestCudaStream.test_add_callback   sG    	 "*Hn5++C01r   c                      t        j                          fd}t        j                         }|j	                  |        j                  j                  d             y )Nc                 H    j                  |       j                          y r   )assertIsNoner   )r   r   argr*   r)   s      r   r!   zCTestCudaStream.test_add_callback_with_default_arg.<locals>.callback#   s    c" r   r"   )r#   r$   r   r   r%   r&   r'   r(   s   `  @r   "test_add_callback_with_default_argz1TestCudaStream.test_add_callback_with_default_arg    sF    "*	! H%++C01r   c                 h   K   t        j                         }|j                          d {    y 7 wr   )r   r   
async_done)r)   r   s     r   test_async_donezTestCudaStream.test_async_done+   s#     !!!s   (202c                    K   dt         dt         f fd}g d}|D cg c]  }t        j                   ||             }}t        j                  |  d {   } j	                  t        j                  ||             y c c}w 7 /w)Nvalue_inreturnc                 ^  K   t        j                         }t        j                  d      t        j                  d      }}| |d d  t        j                  ||      }|j	                  ||       |j                          d {   }j                  ||       |j                         S 7 &w)N   r   )r   r   pinned_array	to_devicecopy_to_hostr2   assertEqualmean)r5   r   h_srch_dstd_arydone_resultr)   s         r   async_cuda_fnz9TestCudaStream.test_parallel_tasks.<locals>.async_cuda_fn2   s     [[]F,,Q/1B1B11E5EE!HNN58EuV4 & 1 1 33K[&1::< 4s   BB-B+'B-)            )floatr	   create_taskgatherr&   npallclose)r)   rC   	values_invtasks
values_outs   `     r   test_parallel_tasksz"TestCudaStream.test_parallel_tasks0   sv     	 % 	 E 	  !	@IJ	1$$]1%56	J">>511
Iz:; K1s   B	"B B	B0B	c                    K   t        j                         }t        d      D cg c]  }|j                          }}t	        j
                  |  d {   }|D ]  }| j                  ||        y c c}w 7 #wNrG   )r   r   ranger2   r	   rJ   r=   )r)   r   _done_awsdoneds         r   test_multiple_async_donez'TestCudaStream.test_multiple_async_doneA   se     16q:AF%%':^^X..AQ'  ;.s   "A8A1A8A6$A8c                 .  K   t        d      D cg c]  }t        j                          }}|D cg c]  }|j                          }}t	        j
                  |  d {   }| j                  t        |      t        |             y c c}w c c}w 7 3wrS   )rT   r   r   r2   r	   rJ   assertSetEqualr   )r)   rU   streamsr   rV   rW   s         r   )test_multiple_async_done_multiple_streamsz8TestCudaStream.test_multiple_async_done_multiple_streamsI   sw     */(3(Q4;;=(36=>gFF%%'g>^^X.. 	CIs7|4 4>.s&   BB	BBBB 4Bc                 (  K   t        j                         }|j                         |j                         }}|j                          | d {    | j	                  |j                                | j	                  |j                                y 7 Cwr   )r   r   r2   cancelr&   	cancelledrW   )r)   r   done1done2s       r   test_cancelled_futurez$TestCudaStream.test_cancelled_futureR   sf     ((*F,=,=,?u)*

% 	s   A
BBABN)__name__
__module____qualname__r+   r0   r   r3   rQ   rY   r]   rc    r   r   r   r      sp    2	2 " " < <  ( ( 5 5 & &r   r   c                   :    e Zd Zej                  ed               Zy)TestFailingStreamc                   K   t        j                         }|j                  d      }|j                  d      }t        j                         }|j                  dd|      j                          |j                         }| j                  t              5  | d {    d d d        | j                  |j                                y 7 ,# 1 sw Y   +xY ww)Nz
            .version 6.5
            .target sm_30
            .address_size 64
            .visible .entry failing_kernel() { trap; }
        failing_kernel)rD   r9   )r   current_contextcreate_module_ptxget_functionr   	configure__call__r2   assertRaises	ExceptionassertIsNotNone	exception)r)   ctxmodulerk   r   rW   s         r   test_failed_streamz$TestFailingStream.test_failed_streame   s      ""$&& (   ,,-=>  tF ;DDF  "y)JJ *T^^-.  *)s0   BCC	CC	(CC		CCN)rd   re   rf   r   skipr   rw   rg   r   r   ri   ri   \   s"     ]]/  /r   ri   __main__)r	   r   r#   numpyrK   numbar   numba.cuda.testingr   r   r   r   r   ri   rd   mainrg   r   r   <module>r~      s         F F	 ?@C&\ C& AC&L ?@/ / A/8 zHMMO r   