
    J/Ph                        d dl mZ d dlmZmZ d dlmZ d dlmZ d dl	m
Z
 d dlmZmZ  G d d          Zd	efd
Ze
d             Z eed          d             Ze
d             Z eedd          d             ZdS )    )types)overloadoverload_method)	signature)	nvvmutils)	intrinsic)
grid_group	GridGroupc                       e Zd ZdZddZdS )r
   z0A cooperative group representing the entire gridreturnNc                      dS )zSynchronize this grid groupN r       M/var/www/html/test/jupyter/venv/lib/python3.11/site-packages/numba/cuda/cg.pysynczGridGroup.sync   s      r   )r   N)__name__
__module____qualname____doc__r   r   r   r   r
   r
   	   s.        ::* * * * * *r   r
   r   c                      t                      S )zGet the current grid group.)r
   r   r   r   	this_gridr      s    ;;r   c                 8    t          t                    }d }||fS )Nc                     |                      t          j        d          }|j        }|                    t          j        |          |f          S )N   )get_constantr   int32modulecallr    declare_cudaCGGetIntrinsicHandle)contextbuildersigargsonemods         r   codegenz_this_grid.<locals>.codegen   sI    ""5;22n||6s;;F  	r   )r   r	   )	typingctxr"   r&   s      r   
_this_gridr(      s*    
J

C   <r   cuda)targetc                      d } | S )Nc                      t                      S N)r(   r   r   r   implz_ol_this_grid.<locals>.impl%   s    ||r   r   )r.   s    r   _ol_this_gridr/   #   s       Kr   c                 D    t          t          j        |          }d }||fS )Nc                     |                      t          j        d          }|j        }|                    t          j        |          g ||R           S )Nr   )r   r   r   r   r   r   declare_cudaCGSynchronize)r    r!   r"   r#   flagsr%   s         r   r&   z!_grid_group_sync.<locals>.codegen/   sQ    $$U[!44n||/44dNENN  	r   )r   r   r   )r'   groupr"   r&   s       r   _grid_group_syncr5   +   s.    
EK
'
'C   <r   r   c                     d }|S )Nc                      t          |           S r-   )r5   )r4   s    r   r.   z!_ol_grid_group_sync.<locals>.impl;   s    &&&r   r   )r4   r.   s     r   _ol_grid_group_syncr8   9   s    ' ' ' Kr   N)
numba.corer   numba.core.extendingr   r   numba.core.typingr   
numba.cudar   numba.cuda.extendingr   numba.cuda.typesr	   r
   GridGroupClassr   r(   r/   r5   r8   r   r   r   <module>r@      sX         : : : : : : : : ' ' ' ' ' '             * * * * * * D D D D D D D D* * * * * * * *9    
 
 
 
 
)F###  $# 
 
 
 777  87  r   