
    Spf                        d dl Z d dlmc mZ d dlmc mc mZ d dlZ	d dl
mZ ddlmZ dD ]&Z	  e j        e dd          Z n# e$ r dZY #w xY werFe                                                                D ]\  ZZ ej        eed	
           ej        ZdedededededefdZd ZdedededededefdZdS )    N)
xla_client   )GpuLibNotLinkedError)z.cudajax_cuda12_pluginz._rnnjaxlib)packageCUDA)platform
input_sizehidden_size
num_layersdropoutbidirectionalcudnn_allow_tf32c                   | j         d         j        }|t          j        k    rt          j                                        }n|t          j        k    rt          j                                        }n|t          j	        k    r<t          j
                            t          j                                                  }n^|t          j        k    r<t          j
                            t          j                                                  }nt          d|           t          j                            | j         d         j        |          }| j        d         j        d         }| j        d         j        d         }t!          ||||||	|
|          \  }}|f}t          j                            |t          j                                                  }| j         d         j        }t          j                            |t          j                                                  }t"          st%                      t"                              ||||||	|
||d         |d         
  
        }t          j                            d          }t-          j        ||j        |j        ||g|||||gt          j                            d          t          j                            d          t          j                            |          t          j                            |d          t          j                            g           	          }|j        d
d         |j        dd
         z   S )z
CuDnn RNN.r   zUnknown output type r          	cudnn_rnnF   )call_target_namehas_side_effectbackend_configapi_versioncalled_computationsN)	avals_outdtypenpfloat32irF32Typegetfloat64F64Type	complex64ComplexType
complex128
ValueErrorRankedTensorTypeshapeavals_in)compute_rnn_workspace_reserve_space_sizes_rnnr   build_rnn_descriptorIntegerTypeget_signlesshloCustomCallOptype
StringAttrBoolAttrIntegerAttr	ArrayAttrresults)ctxinputh_0c_0weightsseq_lengthsr   r   r   r   r   r   	out_dtypeout_typeoutput_type
batch_sizemax_seq_lengthworkspace_size_workspace_shapeworkspace_typereserve_space_shapereserve_space_typeopaquei32_typeouts                             N/var/www/html/nettyfy-visnx/env/lib/python3.11/site-packages/jaxlib/gpu_rnn.pycudnn_rnn_loweringrO   '   s   
 mA$)"*z~~HHBJz~~HHBL  ~!!"*.."2"233HHBM!!~!!"*.."2"233HH
7I77
8
88#''a(8(>II+|A$Q'*<?(+.?+z:~}.0 0.! $%/&**?BJNN<L<LMM.a(.*../B/1z~~/?/?A A	 !

 
  $$Zj%/%24D%4Q%7%8%;	= =& ^((,,(CHch8JKc3-}((55kooe,,]&&v...$$Xq11,**2..	 	 	# 
SbS	CK,	,,    c                     t          j        t          j                            t          j        | t
          j                  t          j                                                            S )N)r   )r4   )	r2   constantr!   DenseElementsAttrr#   r   zerosr    r"   )r+   s    rN   _hlo_zeros_f32rU   [   sX    	
(5

+
+
+"*..2B2B  D D
E 
E ErP   c                |   | j         d         j        d         }| j         d         j        d         }t          ||||||||          \  }}|f}t          j                            |t          j                                                  }| j         d         j        }t          t          d          t          	                    |||||||||d         |d         
  
        }t          j
                            d          }t          | j        d         j                  }t          j        |j        |j        |j        |j        |g|||||||||	||
gt          j                            d          t          j                            d	          t          j                            |          t          j                            |d
          t          j                            g           t          j                            t          j                            dgdg           g                    }|j        dd         S )zCuDnn RNN Backward pass.r   r   r      Nzcuda couldn't be importedr   cudnn_rnn_bwdFr   	   )output_tuple_indicesoperand_indexoperand_tuple_indices)r   r   r   r   r   output_operand_aliasesr   )r,   r+   r-   r!   r*   r#   r"   r.   RuntimeErrorr/   r0   r1   rU   r   r2   r3   r4   r5   r6   r7   r8   OutputOperandAliasr9   )r:   dydhndcnxh0c0wyreserve_spacer?   r   r   r   r   r   r   rC   rD   rE   rF   rG   rH   rI   rK   rL   	zeroed_dwrM   s                               rN   cudnn_rnn_bwd_loweringrj   a   s   
 |A$Q'*<?(+.?+z:~}.0 0.! $%/&**?BJNN<L<LMM.Q-	\
2
3
33$$Zj%/%24D%4Q%7%8%;	= =& ^((,,(S]1-344)vrw8
c32r1a	
; }((99kooe,,]&&v...$$Xq11,**2..\--

 
$
$$%3$& % ( (/ 	 		
 	
 	
#  
SbS	rP   )	importlibjaxlib.mlir.irmlirr!   jaxlib.mlir.dialects.stablehlodialects	stablehlor2   numpyr   r   r   gpu_common_utilsr   cuda_module_nameimport_moduler.   ImportErrorregistrationsitems_name_valueregister_custom_call_targetr-   intboolrO   rU   rj    rP   rN   <module>r~      s                , , , , , , , , , , , ,           2 2 2 2 2 26 
 

"9"&6#=#=#=xPPPD 
E 
   DDD
  ]))++1133 K KmeV*J*5&6JJJJJ.2.\+1-#&1-581-FI1- $1-591- *.1- 1- 1- 1-hE E E(FI((+(9<(GK( +/( CG( ( ( ( ( (s   AAA