+
    :i                         ^ RI HtHtHt ^ RIHt ^ RIHtHt ^ RI	H
t
 ^ RIt^ RIt^ RIt]! R4       ! R R]4      4       t]R	8X  d   ]P                   ! 4        R# R# )
    )cudafloat32int32)NumbaInvalidConfigWarning)CUDATestCaseskip_on_cudasim)ignore_internal_warningsNz#Simulator does not produce lineinfoc                   P   a  ] tR t^
t o R tR tR tR tR tR t	R t
R tR	tV tR
# )TestCudaLineInfoc                2    R p\         P                  ! V4      # )z \.loc\s+[0-9]+\s+[0-9]+\s+[0-9]+)recompile)selfpats   & c/var/www/html/photoedit/myenv/lib/python3.14/site-packages/numba/cuda/tests/cudapy/test_lineinfo.py_loc_directive_regex%TestCudaLineInfo._loc_directive_regex   s     	 zz#    c                   VP                  V4       VP                  V4      pVP                  V4      pV'       d   V P                  MV P                  pR p\
        P                   ! V4      P                  V4      pV! WR7       Rp\
        P                   ! V4      P                  V4      pV P	                  WR7       Rp\
        P                   ! V4      P                  V4      pV! WR7       V P                  4       P                  V4       V! WR7       Rp\
        P                   ! V4      P                  V4      pV P	                  WR7       R# )z5!DICompileUnit\(.*emissionKind:\s+DebugDirectivesOnly)msgz+!DICompileUnit\(.*emissionKind:\s+FullDebugz&\.file\s+[0-9]+\s+".*test_lineinfo.py"z\.section\s+\.debug_infoN)r   inspect_llvminspect_asmassertIsNotNoneassertIsNoner   searchr   )	r   fnsigexpectllvmptxassertfnr   matchs	   &&&&     r   _checkTestCudaLineInfo._check   s!   


3s#nnS!+14''t7H7H
# 	 

3&&t,  	
 

3&&t,%)
$ 	
 

3&&s+  	!!#**3/ 
 	 

3&&s+%)r   c                    \         P                  ! R R7      R 4       pV P                  V\        R,          3R R7       R# )Flineinfoc                     ^V ^ &   R#    N xs   &r   foo5TestCudaLineInfo.test_no_lineinfo_in_asm.<locals>.fooK       AaDr   NNNr   r   Nr   jitr#   r   r   r.   s   & r   test_no_lineinfo_in_asm(TestCudaLineInfo.test_no_lineinfo_in_asmJ   s7    	5	!	 
"	 	CeAh[7r   c                    \         P                  ! R R7      R 4       pV P                  V\        R,          3R R7       R# )Tr&   c                     ^V ^ &   R# r)   r+   r,   s   &r   r.   2TestCudaLineInfo.test_lineinfo_in_asm.<locals>.fooR   r0   r   r1   r2   Nr3   r5   s   & r   test_lineinfo_in_asm%TestCudaLineInfo.test_lineinfo_in_asmQ   s7    	4	 	 
!	 	CeAh[6r   c                    \         R ,          \         R ,          3p\        P                  ! VRR7      R 4       pVP                  V4      pV P	                  RV4       R# ):NNr*   Tr&   c                 >    V ^ ;;,          V^ ,          ,          uu&   R# r   Nr+   )r-   ys   &&r   divide_kernelKTestCudaLineInfo.test_lineinfo_maintains_error_model.<locals>.divide_kernel[   s    aDAaDLDr   z	ret i32 1N)r   r   r4   r   assertNotIn)r   r   rA   r   s   &   r   #test_lineinfo_maintains_error_model4TestCudaLineInfo.test_lineinfo_maintains_error_modelX   sU    s|WS\*	#	%	 
&	 ))#. 	d+r   c                   a \         P                  R  4       o\         P                  V3R l4       p\        R,          3pV P                  WRR7       R# )c                 0    V ^ ;;,          ^,          uu&   R# r?   r+   r,   s   &r   calleeDTestCudaLineInfo.test_no_lineinfo_in_device_function.<locals>.calleei       aDAIDr   c                 "   < ^V ^ &   S! V 4       R# r)   r+   r-   rH   s   &r   callerDTestCudaLineInfo.test_no_lineinfo_in_device_function.<locals>.callerm       AaD1Ir   r1   Fr2   N)r   r4   r   r#   )r   rM   r   rH   s   &  @r   #test_no_lineinfo_in_device_function4TestCudaLineInfo.test_no_lineinfo_in_device_functiong   sN    		 
	 
	 
	 QxkFE2r   c                  a \         P                  ! R R7      R 4       o\         P                  ! R R7      V3R l4       p\        R,          3pV P                  WR R7       VP	                  V4      pVP                  4       p\        P                  ! R4      pV F,  pVP                  V4      f   K  V P                  RV 24       K.  	  V P                  4       pR	pV F#  pVP                  V4      f   K  R
V9   g   K!  R p M	  V'       g   V P                  RV 24       VP                  V4      p	^ p
V	P                  4        F  pRV9   g   K  V
^,          p
K  	  ^pV P                  WRV RV
 24       R# )Tr&   c                 0    V ^ ;;,          ^,          uu&   R# r?   r+   r,   s   &r   rH   ATestCudaLineInfo.test_lineinfo_in_device_function.<locals>.calleey   rJ   r   c                 "   < ^V ^ &   S! V 4       R# r)   r+   rL   s   &r   rM   ATestCudaLineInfo.test_lineinfo_in_device_function.<locals>.caller}   rO   r   r1   r2   z^\.weak\s+\.funcNzFound device function in PTX:

F
inlined_atz1No .loc directive with inlined_at info foundin:

zdistinct !DISubprogramz
"Expected z DISubprograms; got )r   r4   r   r#   r   
splitlinesr   r   r"   failr   r   r   assertEqual)r   rM   r   r    ptxlinesdevfn_startlineloc_directivefoundr   subprogramsexpected_subprogramsrH   s   &           @r    test_lineinfo_in_device_function1TestCudaLineInfo.test_lineinfo_in_device_functionu   s    
4	 	 
!	 
4	 	 
!	 QxkFD1   %>>#
 jj!45D  &2		=cUCD  113D##D)54' E	  II   #u& ' ""3'OO%D'4/q  &  !%&:%; <  +}.	/r   c                   \         P                  ! R R7      ;_uu_ 4       p\        4        \        P                  ! R R RR7      R 4       pRRR4       V P                  \        X4      ^4       V P                  V^ ,          P                  \        4       V P                  R\        V^ ,          P                  4      4       R#   + '       g   i     L; i)T)recordF)debugr'   optc                      R # )Nr+   r+   r   r   f;TestCudaLineInfo.test_debug_and_lineinfo_warning.<locals>.f   s    r   Nz)debug and lineinfo are mutually exclusive)warningscatch_warningsr	   r   r4   rZ   lencategoryr   assertInstrmessage)r   wri   s   &  r   test_debug_and_lineinfo_warning0TestCudaLineInfo.test_debug_and_lineinfo_warning   s    $$D11Q$& XXD4U; < 2 	Q#1(ABA!A$,,'	) 21s   *CC	r+   N)__name__
__module____qualname____firstlineno__r   r#   r6   r;   rD   rP   rb   rs   __static_attributes____classdictcell__)__classdict__s   @r   r   r   
   s4     	1*f87,3?/B) )r   r   __main__)numbar   r   r   numba.core.errorsr   numba.cuda.testingr   r   numba.tests.supportr	   r   unittestrk   r   ru   mainr+   r   r   <module>r      sZ    & & 7 < 8 	   67x)| x) 8x)v zMMO r   