o i9:j!@s5UddlZddlZddlZddlZddlZddlmZmZddlm Z ddl m Z ddlm Z m Z ddlmZmZddlmZmZddlZddlZddlmZddlmZddlmmZddlmZm Z m!Z!dd l"m#Z#dd l$m%Z%dd lm&Z&m'Z'm(Z(m)Z)m*Z*dd l+m,Z,m-Z-m.Z.m/Z/dd l0m1Z2ddl3m4Z4ej5j6Z6gZ7e8e9e:d<ej;j  ddedej?de@de@fddZAe eAej?jBddZCe eAej?jBdZDe eAej?jBddZEe eAej?jFdZGde!deHde!fd d!ZIe#e=jJe/d"eDd#e!d$e!fd%d&ZJe#e=jKe/d"eDd#e!d$e!fd'd(ZKe#e=jLe/d"eDd#e!de!d)eMd*eMfd+d,ZLe#e=jNe/d"eDd-e!d.eMd/eMd0eMd1e@d2e!f d3d4ZNe#e=jOjPgd5d6ZQe#e=jOj!gd7e!fd8d9ZRe#e=jSe/eDd:e!de!fd;d<ZSe#e=jTe/d"eDd-e!d:e!fd=d>ZTe#e=jUe/d"d-e!d:e!d?eMd@eMfdAdBZUe#e=jVe/eDd:e!de!fdCdDZVe#e=jWe/eDd-e!d:e!de!fdEdFZWe#e=jXe/d"d-e!d:e!d*eMfdGdHZXe#e=jYe/d"eDd-e!d:e!dIeMdJe@fdKdLZYe#e=jZe/d"eDddNe!d:e!dOe9fdPdQZZe#e=j[eDd-e!dRe!fdSdTZ[e#e=j\e/eDd:e!de!fdUdVZ\e#e=j]e/d"eDd-e!d:e!de!fdWdXZ]e#e=j^d:e!dYe!de!fdZd[Z^e#e=j_d-e!d:e!dYe!de`e!e!ffd\d]Z_e#e=jae/eDd-e!d:e!d^e!d_eMd`eMdae@dJe@de!fdbdcZae#e=jbe/d"eDd-e!d:e!dde!de!fdedfZbe#e=jce/d:e!dge!de!fdhdiZcdje!dkeHfdldmZddnejefdodpZfe#e=jge/eDe>jhjifd:e!dqe!dkeHde!fdrdsZge#e=jje/d"eDd-e!dRe!dqe!dkeHfdtduZje#e=jkddvdwZle#e=jme/eDe>jhjidxfd:e!dqe!dkeHd)eMfdydzZme#e=jnjoeDd-e!d:e!dqe!dkeHd)eMf d{d|Zne#e=jnjpeDd-e!d:e!dqe!dkeHd)eMd"e!f d}d~Zqe#e=jrjoeDd-e!d:e!dqe!dkeHdeMf ddZre#e=jrjseDd-e!d:e!dqe!dkeHdeMd"e!f ddZtd-e!d:e!dqe!dYe!dBdkeHdeHde!de!fddZue#e=jve/d"eDd-e!d:e!deHde!fddZve#e=jwe/d"d-e!d:e!dqe!dYe!dBdkeHdeHde!de!fddZwe#e=jxe/d"d-e!d:e!dqe!dYe!dBdkeHdeHde!de!fddZxe#e=jye/eDde>jhjifd:e!dqe!dYe!dBdkeHde!f ddZye#e=jze/d"eDde>jhjifd-e!d:e!dqe!dYe!dBdkeHde!f ddZze#e=j{e/eDe>jhjifdRe!dqe!dkeHde!fddZ{e#e=j|e/d"eDe>jhjifd-e!d:e!dqe!dkeHde!f ddZ|e#e=j}e/ddRe!dge!deMfddZ}e#e=j~e/de!de!de!fddZ~e#e=je/d-e!de8eHdeHdeHdeHdeHf ddZe#e=jj!   dd:e!deHdeHdBdeHdBdeHf ddZde!deHdeHdBdeHdBde`eHeHff ddZe#e=je/   ddRe!de!deHdeHdBdeHdBdeHf ddZe#e=je/d-e!de8eHdeHdeHfddZe#e=je/d-e!de8eHdeHdeHdeHf ddZd-e!d"e!dejefddZe#e=je/d"eCd-e!de!deHdejefddZe#e=je/eCd-e!de!deHdejefddZddZe#e=je/dRe!de8eHde8eHde8eHde8eHde!f ddÄZe#e=je/eDdRe!de8eHde8eHde8eHde8eHde8eHde!fddƄZe#e=je/d-e!de!d/eMfddɄZe#e=je/dNe!de8eHdeHdeHdeHde!f dd΄Ze#e=jjoeD dd-e!d:e!deMdBde!fddфZe#e=je=jjoe6je=jjoe6jdRe!deMde@dBfddԄZe#e=je/ddփdRe!deMde@dBfdd؄Ze#e=je/de!deHde@fddۄZe#e=je/dd܍de!deHde@fddބZe#e=je/  ddYe!de!deHde@de@de!f ddZe#e=je/d-e!de!deHdeHde@f ddZde8eHfddZde8e!deHdeHde8e!fddZde8e!fddZde8e!deHfddZde8e!deHdeHfddZe#e=jjoe=jjsg dde8e!deHdeHde!dBde!f ddZe#e=jjoe=jjsg dd:e!de8eHdeHde8e!dBde8e!dBf ddZe#e=jj!ddRe!deHdeHde`e!dffddZe#e=jjo ddRe!de8eHdeHde`e!dffddZe#e=jj!dd:e!deHdeHde`e!dffddZe=jje6j dd:e!de!deHde`e!dffddZe#e=je/dd܍eDdd:e!de!de!d)eHd.eHf dd Ze#e=je/eD dd:e!de!de!d)eHd.eHd e@f d d Ze#e=je/dd܍eDdd:e!de!d e!d)eHd.eHf ddZe#e=jjoeDd-e!dRe!de!de!de!dBdeHdeHdeHdeHde8e@de`e!dBe!dBe!dBffddZe#e=jjsd-e!dRe!de!de!de!dBdeHdeHdeHdeHde8e@dej!dej!dej!de`e!dBe!dBe!dBffddZde!dBde!dBfddZe#e=jjode!dRe!d e8eHde!de!dYe!dBd!e!dBde8e@de`e!dBe!dBe!dBffd"d#Ze#e=jjsde!dRe!d e8eHde!de!dYe!dBd!e!dBde8e@dej!dej!dej!de`e!dBe!dBe!dBffd$d%Ze#e=jjodRe!d e8eHdYe!dBdeMdBde`e!e!ff d&d'Ze#e=jjode!dRe!d e8eHde!dYe!dBde8e@de`e!dBe!dBffd(d)ZdRe!dYe!dBd!e!dBd*e!dBd+e!dBdae@d,eMdeMd-e@de`e!e!e!e!dBe!dBffd.d/Ze#e=je/dd0d1dRe!dYe!dBd!e!dBd*e!dBd+e!dBdae@d,eMdeMde`e!e!e!ffd2d3Ze=jjoe6je=jjoe6jdRe!dYe!dBd!e!dBd*e!dBd+e!dBdae@d,eMdeMde`e!e!e!ffd4d5Ze=jjoe6jdde8e!fd6d7Ze#e=jjodRe!dYe!dBd!e!dBd*e!d+e!d,eMdeMde`e!e!e!ffd8d9Ze#e=jjodRe!dYe!dBd!e!dBd*e!d+e!dae@d,eMdeMde`e!e!e!ffd:d;Ze#e=jjdRe!dYe!dBd!e!dBdae@d,eMdeMde`e!e!e!ffd<d=Ze#e=jjodRe!dYe!dBd!e!dBd*e!d+e!dae@d,eMdeMde`e!e!e!e!e!ffd>d?ZdRe!dYe!dBd!e!dBd*e!d+e!deMdae@de!fd@dAZe#e=jjodRe!dYe!dBd!e!dBd*e!d+e!d,eMdeMde`e!e!e!e!ffdBdCZe#e=jjodRe!dYe!dBd!e!dBd*e!d+e!d,eMdeMde`e!e!e!e!e!e!ffdDdEZe#e=jjodRe!dYe!dBd!e!dBd*e!d+e!d,eMdeMde`e!e!e!e!ffdFdGZe#e=je/ddփeDddHdIZe#e=je/dddddddJde!e'BdnejedBdKejdBdLe@dMe@dNejdBf dOdPZe#e=je=je=jge/dQdRZe=jjoe6je#e=jƃe/dd֐ddSdRe!dYe!d!e!dBd*e!dBd+e!dBdae@dTeMdUeMfdVdWZƐdXdYZe#e=jjode!dRe!dYe!dBd*e!dBd+e!dBd0e!dBd1e!dBde@deMde8e@dZe!de`e!e!dBe!dBffd[d\Ze#e=jjode!dRe!dYe!dBd*e!dBd+e!dBd0e!dBd1e!dBde@deMde8e@de`e!e!dBe!dBffd]d^Ze#e=jjsde!dRe!dYe!dBd*e!dBd+e!dBd0e!dBd1e!dBde@deMde8e@dej!dej!dej!de`e!e!dBe!dBffd_d`Ze#e=j˃e/dd֐ddRe!d-e!dYe!d*e!dBd+e!dBd0e!dBdae!dBdUeMfdbdcZe#e=j̃e/dd֐ddRe!d-e!dYe!d*e!dBd+e!dBd0e!dBdae!dBdUeMdde!fdedfZe#e=j̓e/eDdRe!de`eHeHffdgdhZd:e)de)de8eHdeHfdidjZe#e=jЃe/d:e)de)de8eHfdkdlZe#e=jуe/dRe)de)de8eHde8eHde8eHf dmdnZe#e=j҃ddode)deHde)dpe)d.e'f dqdrZe#e=jӃe/ddode)deHde)dpe)d.e'f dsdtZddode)deHde)dpe)due@d.e'f dvdwZe#e=jjoe=jjoe6jddydzZe#e=jփde)deHde)dpe)fd{d|Ze#e=j׃e/de)deHde)dpe)fd}d~Zde)deHde)dpe)due@f ddZe#e=jكe/dddeDd:e!de`e!e!ffddZe#e=jڃe/ x x dde!de@eHBeMBde@eHBeMBdejdBfddZe#e=j܃dddZܐddZݐddZe#e=jje#e=jje#e=jje=jjࠐe6je=jjࠐe6je=jjࠐe6je=jjࠐe6je=jjࠐe6je=jjࠐe6jdRe!de8eHdBde8eMdBde!fddZe#e=jje#e=jje#e=jje=jjࠐe6je=jjࠐe6je=jjࠐe6je=jjࠐe6je=jjࠐe6je=jjࠐe6jdRe!de8eHdBde8eMdBde!fddZdddZe#e=jjoe=jjsge=jjoe6je=jjoe6je/ddd ddRe!de8eHdeMdBde!fddZe#e=jjoe=jjsge=jjoe6je=jjoe6je/ddd ddRe!de8eHdeMdBde!fddZe#e=jjoe=jjsge=jjoe6je=jjoe6je/ddd  ddRe!de8eHdeMdBdeMdBde!f ddZe#e=jjoe=jjsge=jjoe6je=jjoe6je/ddd  ddRe!de8eHdeMdBdeMdBde!f ddZe#e=jjoe=jjsge=jjoe6je=jjoe6je/ddd   ddRe!de8eHdeMdBdeMdBdeMdBde!f ddZe#e=jjoe=jjsge=jjoe6je=jjoe6je/ddd   ddRe!de8eHdeMdBdeMdBdeMdBde!f ddZeD ddRe!de8eHde8eMdBde@de!f ddZddZddZddZddZ dddZddZddZdddZdddZddZe#e=jje=jje6je=jje6jddZe#e=jje=jje6je=jje6jddZe#e=jje=jje6je=jje6jdd„Ze#e=jje=jje6je=jje6jdÐdĄZdŐdƄZddǐdȄZddɐdʄZdːd̄Ze#e=jje=jje6je=jje6jd͐d΄Ze#e=jje=jje6je=jje6jdϐdЄZdѐd҄ZdӐdԄZe#e=jje=jje6je=jje6jdՐdքZe#e=jje=jje6je=jje6jdאd؄Ze#e=j je=j jࠐe6je=j jࠐe6jdِdڄZ e#e=j je=j jࠐe6je=j jࠐe6jdېd܄Z e#e=j je=j jࠐe6je=j jࠐe6jdݐdބZe#e=jje#e=jje=jjࠐe6je=jjࠐe6je=jjࠐe6je=jjࠐe6je=jjࠐe6je=jjࠐe6jdߐdZe#e=jjoe=jjsge/ ddRe!de8eHde@deMdBde!f ddZe#e=jjoe=jjsge=jjoe6je/  ddRe!de8eHde@deMdBdeMdBde!f ddZe#e=jjoe=jjsge/   ddRe!de8eHde@deMdBdeMdBdeMdBde!fddZdddZddZdee!dee!de!de!fddZde*de!fddZeDdRe!de8eHde@de8eMdBde!f ddZe#e=jjode!de!de@fddZe#e=je=jge/ddZe#e=jgddZe#e=jgdddZe#e=jgddZe#e=jgddZd:e!dqe!dYe!dBdkeHdeHde`e!e!ff ddZe#e=j e/ddd:e!dqe!dYe!dBdkeHdeHde`e!e!ff ddZ e#e=j!e/ddd:e!dqe!dYe!dBdkeHdeHde`e!e!ff ddZ!de!deMde!fd d Z"de!deMde!fd d Z#d e!de*fddZ$de*de!de!fddZ%dee!de!fddZ&deHde@dnejedKejfddZ'de!deHdeHde@fddZ(de!deHdeHdeHde@f dd Z)de!de8eHde@fd!d"Z*de!de8eHde@fd#d$Z+e#e=j,e/eDde!de8eHde@fd%d&Z,  dde!d'e!d(eHd)eHde@d*e@de!fd+d,Z-e#e=j.e/eD dde!d'e!d(eHd)eHde@de!f d-d.Z.e#e=j/e/dd܍eDd/d0Z/e#e=j0e/dde>jhjifd1d2Z0d3ej!d4ej!d5e@de@fd6d7Z1e=j2joe6je=j2jse6je/dd8dd9d:d;Z2e#e=j3joe=j3jsge=j3joe6je/eD  ddRe!de`eHeHfde@dd?Z4e#e=j3je=j3jࠐe6je=j3jࠐe6je/eD dde!de`eHeHfdBde@de`eMeMfdBde!f d@dAZ5e#e=j6e#e=j7e#e=j8eDe/de!de`eHdfde!fdBdCZ9e#e=j:e#e=j;e#e=j<eDe/de!de`eHdfde!fdDdEZ=de!de`eHdfdFeeHeHeHge!fde!fdGdHZ>e#e=j?e#e=j@e#e=jAe/d"dIdJZBe#e=jCe/dKdLdddMdNdOZCe#e=jDe/dddPdQdRZDe#e=jEjoe=jEjsge/dejFdddSde'dnejedBdTejGdKejdBdLe@f dUdVZHe#e=jEjIgdejFdddSde'de'dnejedBdTejGdKejdBdLe@f dWdXZJe#e%dYdZZKe#e=jLe=jLjoe6je/ddde>jhjifdRe!dqe!de'd[e'dYe!dBdkeHde!fd\d]ZLe#e=jMe=jMjoe6je/dd^dRe!dqe!dkeHde`e!e!ffd_d`ZMe#e=jNjo x ddddadbe!dce!d7e!ddeMdee@dfe!dBd/eMdBde`e!e!ffdgdhZOdidjZPe#e=jQge/dd܍eDddkdlZQe#e=jRe/dmdnZRe#e=jSdodpZSe#e=jTjoe=jTjsgdddqd:e!dnejedBde!dBde!fdrdsZUe#e=jVjoe=jVjWgdd:e!deHdBfdtduZXe#ejEs z-type_casts..inner..type_promotion_kindct|tr |S|SNr-rtor0computation_dtyper'r( increase_precO  z0type_casts..inner..increase_preccr4r5r6r8) result_dtyper'r( decrease_precUr<z0type_casts..inner..decrease_prec) rtorchtypes_Numberpytreearg_tree_leavesutilselementwise_dtypesr)argskwargs flat_argsr;r>rr+r)r,r*)r1r:r=r(inner@s     ztype_casts..inner) functoolswraps)r)r*r+r,rKr'rJr( type_casts:s rNT)r*r+)r*)r*r,r0dimreturncCs$t||D]}|d}q|SN)rangerO unsqueeze)r0rO_r'r'r(_unsqueeze_to_dimws rV grad_inputout_gradycCs|d||SNr conj_physicalrXrYr'r'r( tanh_backward}r^cCs||d|SrZr[r]r'r'r(sigmoid_backwardr_r`beta thresholdcCs.||}t|||k||||dSN?)expr?where)rXr0rarbzr'r'r(softplus_backwards "rh grad_outputalphascale input_scale is_resultself_or_resultc Csb||}|}|}|rt|dk||||||St|dk|||t||||SNr)r?rfre) rirjrkrlrmrnnegcoefposcoef negiptcoefr'r'r( elu_backwards rscC t||Sr5)r? full_likeselfvaluer'r'r( fill_scalar ryrxcs(tdkfddt|S)NrcsddS)Nz@fill only supports 0-dimension value tensor but got tensor with z dimensionsrOr'rxr'r(zfill_tensor..)r?_checkrOatencopyrvr'r|r( fill_tensors    rrwcCs tjtj|ddddddSNrminmaxr?clamprwr'r'r( hardsigmoids rcCst|dk|dk@|ddS)Ngg@gUUUUUU?r?rfrirwr'r'r(hardsigmoid_backwards rmin_valmax_valcCst||k||kBd|S)Nrr)rirwrrr'r'r(hardtanh_backwardsrcCs$|tjtj|ddddddSrrrr'r'r( hardswishs$rc Cs,t|dkdt|dk||dd|S)Nrr?rrr'r'r(hardswish_backwards rcCst||kd|Sror)rirwrbr'r'r(threshold_backwardrnegative_slopeself_is_resultcCst|dk|||Sror)rirwrrr'r'r(leaky_relu_backwardsrnonegrad approximatecCsd}d}d}|dkrO||d}d}||}||} |||| } t| } d|} d| } d| }d| | }|dd||}| ||}|||S|}||d}ddt||}|t||d }||||S) Ng;f?g;f?gmBP ?tanhrgHm?rr)r?rerfre)rrwrM_SQRT2 M_SQRT1_2 M_2_SQRTPIkBetakKappax_sqx_cuberK tanh_innerleftrightleft_derivativetanh_derivativeinner_derivativeright_derivativekAlphacdfpdfr'r'r( gelu_backwards,      rinputcCs:tt|}t|}||d||}|||SrZ)r?rFsoftplussigmoid)ririnput_tanh_softplus input_sigmoidoutr'r'r( mish_backwards  rcCs|t|Sr5)r?rrr'r'r(silu"srcCs,ddt| }||d|d|SrZ)r?re)rirwrr'r'r( silu_backward)srweightcCst|dk|||Sror)rwrr'r'r( _prelu_kernel1srcCs4t|dk|||}t|dkd||}||fS)Nrrr)rirwr input_grad weight_gradr'r'r(_prelu_kernel_backward6srnoiseloweruppertrainingcCs*|r||S||d}t||||SNr )mulrr)rirwrrrrrrr'r'r(rrelu_with_noise_backwardAs   rbuffercCs^|dk}t|dd}t|dd}|dkr|ntt| }||||d|S)NrrrR)r?rfnumelreabs)rirwr in_negative max_derivsignrgr'r'r(log_sigmoid_backwardVs "rothercCsDt|js t|jrtjn|j}|jdd|d}|t||S)Nr'@dtype)rDis_integer_dtyperis_boolean_dtyper?float32new_fullpow)rwr two_dtype two_tensorr'r'r(ldexpcs  rloss reductioncCs0|tjjkr t|S|tjjkrt|S|Sr5)rr%rxr?meanr&sum)rrr'r'r(apply_loss_reductionos    rrcCs4|tjkrtjS|tjkrtjS|tjkrtjSdSr5)r? complex32float16 complex64r complex128float64rr'r'r( to_real_dtypexs   rtargetcCs||d}t||Sr)r)rwrrrr'r'r(mse_losss  rcCs,|tjjkr d|nd}||||S)Nr)rr%rxr)rirrrnormr'r'r(mse_loss_backwardsrcCsFtj|||d}|td}tj||dd}t|}t|||S)N)rOrz-infTrOkeepdim)r?softmaxeqfloatall zeros_likerf)rwrOrrmasked masked_rowszerosr'r'r( safe_softmaxs  rrdcCs<||}t||kd|d||d|}t||S)Nrr )rr?rfr)rwrrrarr'r'r(smooth_l1_losss & rc CsZ|tjjkr d|nd}||}t|}||}t||k||||t|Src)rr%rxrr?rrfr) rirwrrrarr0abs_x norm_gradr'r'r(smooth_l1_loss_backwards   rcC*t|||||}t||jt||ddSNT copy_fromcopy_to exact_dtype)rrshaper)rirwrrrarWresultr'r'r(smooth_l1_loss_backward_out rdeltac Cs`|tjjkr d|nd}||}t|| k| ||t||k||||||Src)rr%rxrr?rf)rirwrrrrr0r'r'r(huber_loss_backwards  rcCrr)rrrr)rirwrrrrWrr'r'r(huber_loss_backward_outrr  ignore_index total_weightc Cs|dkrdnd}|tjjkr||}|dkr$|dkr$|d}||}t||k|d}t|} t| ||d} | |krMdkrTnn||}|dursddt |D} |j d| |<| | }||}t||k|d}| |S)Nr rrgcSsg|]}dqSrr'r/rUr'r'r(r2 r~z&_nll_loss_backward..) rOrr%rxrTr?rfrscatterrSrreshape) rirwrrrr r  channel_dim safe_targetrW new_shaper'r'r(_nll_loss_backwards$      rc Cs|dkr tdt||}||}|ddkr'td|d||d}||d|}||||}t|}d||||} ||}tj|| g|dS)Nrz*glu does not support 0-dimensional tensorsr z.Halving dimension must be even, but dimension z is size rdr{) rOAssertionErrorrDcanonicalize_dimsizenarrowr?rcat) rirwrOwrap_dimnIn inputSize firstHalf secondHalfgradInputFirstHalfgradInputSecondHalfr'r'r( glu_backwards    r cCsd|kr dksntd|d|dkr'td|d|dko2|dk}|sL|jd|jdksLtd|jd|jd |dkr`td |jd |d |durq||jd krqtd|tjjkr|dkr|dkr|jd|jdkstd|jdd|d|jdn|dkr|dkstd|jt|||||||S)Nrr %input tensor should be 1D or 2D, got DrA0D or 1D target tensor expected, multi-target not supported, got size mismatch (got input: , target: ):expected total_weight to be a single element tensor, got: z (z elements)rRz.clamp_wrapr)rint)r0rOrIrJrgr'rer(_normalize_start_end%s risrcc CsBt|j|}|j|}t||||\}}t|j}|||d|||<||}|dkr;||kr;|dkr;|Sdg|}t j ||j d} | ||||<t j ||j t j d} |dkrht | | |k} ||krtt | | |k} |dkrt | | ||dk} dg|} d| |<| | } t| t|| |d|S)NrrdevicerlrrR)rDrr[rrirSexpandclonerOr?arangerlonesbool logical_andviewrrf_unsafe_masked_index) rrjrOrIrJrKrfsrc_sizeindicesidxmask mask_shaper'r'r(rM<s,    rMindexcCs||}t||||Sr5)rLr?select_scatter)rirHrOr{rWr'r'r(select_backwardes r}offsetdim1dim2cCs||}t|||||Sr5)rLr?diagonal_scatter)rirHr~rrrWr'r'r(diagonal_backwardls r input_dtypecCs|j|kr ||}|Sr5)rr7)rirWrr'r'r(_cast_grad_to_input_dtypeus  routputcCs0||}||tj||dd}t|||SNTr)r?rr contiguous)rirrOrnew_grad_outputrWr'r'r(_softmax_backward_data}s  rcCs*|t|tj||dd}t|||Sr)r?rerr)rirrOrrWr'r'r(_log_softmax_backward_datas rc CsZ||d||d}ttjtj|d}|d||d}|d|||d} || S)z/Utility function to implement im2col and col2imr rrrlrrR)rr?rpint64rT) input_dkernel_d dilation_d padding_dstride_drlblocks_d arange_kwblocks_d_indices kernel_gridr'r'r( _im2col_col2im_indices_along_dims r kernel_sizedilationpaddingrTc s&ttdkddttdkddttdkddttdkddddd }|d |d |d d d|d|jt}t|dvodtddddDfddtddtddDttddDfdd|dk}|s|d}|j\}} } } \} } \}}\}}\}}t| |||| |j }t| |||| |j }t |||||f}|dd}|dddd||f}| dddddd}| d}| d}||| ||||}|s|d}|S) Nr cSdS)Nz"im2col(): only 2D kernel supportedr'r'r'r'r(r}zim2col..cSr)Nz$im2col(): only 2D dilation supportedr'r'r'r'r(r}rcSr)Nz#im2col(): only 2D padding supportedr'r'r'r'r(r}rcSr)Nz"im2col(): only 2D stride supportedr'r'r'r'r(r}rTcB|r tddDntddD}t|fdddS)Ncs|]}|dkVqdSrNr'r/r4r'r'r( z1im2col..check_positive..cs|]}|dkVqdSrr'rr'r'r(rrcdSNz& should be greater than zero, but got r'r'param param_namer'r(r}z0im2col..check_positive..rr?rrrstrictcondr'rr(check_positive(zim2col..check_positiverrrFrrTrr*cs|]}|dkVqdSrr'r/dr'r'r(rrzim2col..rcdtS)NzmExpected 3D or 4D (batch mode) tensor for input with possible 0 batch size and non-zero dimensions, but got: tupler'rr'r(r}css>|]\}}}}}d|d|||dd|VqdS)rr Nr'r/rpaddilkerstr'r'r(rs  " r:csrrr')r/cr'r'r(rrc s6dtdddddddd S) Nz!Given an input with spatial size r:, kernel_size= , dilation= , padding= , stride=z9, the calculated shape of the array of sliding blocks is z*, but its components must be at least one.rr'rr output_sizerrrTr'r(r}sr*rrRrrT)r?rr`rrrziprTrrlrrpermuterrsqueeze)rrrrrTrr[ batched_input batch_dimrinput_hinput_wstride_hstride_w padding_h padding_w dilation_h dilation_wkernel_hkernel_wblocks_row_indicesblocks_col_indices padded_inputrnum_blocks_rownum_blocks_colr'rr(im2colsd           rrc sttdkddttdkddttdkddttdkddttdkddd$d d }|d |d |d dd|d|d|jt}t|dvoutddddDfdddd}td|dkfddddtD} | d| dtdkfddtdkfdd|dk} | s|d}|j\} } \} }\}}\}}\}}|dd|g| }|dddd dd!}t | |||| |j }t |d }t | |||||j }d"dtD}| ddt g|}dd||f}tj|||dd#}t|| | | | f}| sf|d}|S)%Nr cSr)Nzonly 2D output_size supportedr'r'r'r'r(r} rzcol2im..cSr)Nzonly 2D kernel supportedr'r'r'r'r(r} rcSr)Nzonly 2D dilation supportedr'r'r'r'r(r} rcSr)Nzonly 2D padding supportedr'r'r'r'r(r}rcSr)Nzonly 2D stride supportedr'r'r'r'r(r}rTcr)Ncsrrr'rr'r'r(rrz1col2im..check_positive..csrrr'rr'r'r(rrcrrr'r'rr'r(r}rz0col2im..check_positive..rrr'rr(rrzcol2im..check_positiverrrFrrTr)r rcsrrr'rr'r'r(r rzcol2im..r:cr)NzmExpected 2D or 3D (batch mode) tensor for input with possible 0 batch size and non-zero dimensions, but got: rr'rr'r(r}!rrrcsdddS)Nz|Expected size of input's first non-batch dimension to be divisible by the product of kernel_size, but got input.shape[-2] = r:z and kernel_size=r'r')rrr'r(r}'s cSs:g|]\}}}}}d|d|||dd|qSrr r'rr'r'r(r2+s "zcol2im..rRc4ddddddddd S NzGiven output_size=rrrrz , expected input.size(-1) to be but got rR.r'r'LrrrrrrTr'r(r}4crrr'r'rr'r(r}:rrr*rcSsg|] \}}|d|qSr r')r/or4r'r'r(r2V accumulater)r?rr`rrrrTrrrrlrVrLprodr_unsafe_index_putrrr)rrrrrrTrr[prod_kernel_sizecolrout_hout_wrrrrrrrr indices_row indices_coloutput_padded_sizerrxr'rr(col2ims            "   rrycCs$||||jt|d}|SNr8)type_asrorDr)riryrkrIr'r'r(native_dropout_backwardcs r input_size dimensionrc Cst|dkr t|dStt||}tj|||jtjd}|d|| }| d|d ||d}| |}d||f}t j |||ddS)NrrmrRrr5Tr)r`r? squeeze_copyrDrrprlint32unfoldflattenmovedimrLrrr) rrrrrKrOrxrWr{r'r'r(unfold_backwardrs   repsc Cst|dur|}d|}tt||k||k||d|dStt|dk|dk||d||dtdS)Nrdrr'nan)r?rfrsrr)rirwrlohir'r'r(logit_backwardsrtraincCs&|r|dkrt|||dS|Sro)rnative_dropoutro)rr4rr'r'r(dropouts rout0out1cCs|r6|dkr6|dkrt|tj|tjdfS|jjstdt||k}||tdd|}||fS|tj|tjdfS)Nrrrz?result type Float can't be cast to the desired output type Longrd) r?rrrris_floating_pointrR rand_likerr;)rr4r bool_maskresr'r'r(rs r half_to_floatc Csddlm}|}|r|jtjkrtd|jdtj|tj j d\}}| |}|| dkr9t |}ntj||dd}t ||}|tj||dd}|sY| |}|SNrguard_or_falsez%half_to_float is True but x.dtype is z, expected torch.halfr3T)r)rQrrrr?halfrrDrEELEMENTWISE_TYPE_PROMOTION_KINDDEFAULTr7rreamaxr) r0rOr rr:r= unnormalizedx_maxrr'r'r(_softmaxs&      r)rc Csddlm}|}|r|jtjkrtd|jdtj|tj j d\}}| |}|| dkr6|}n tj ||dd}||}ttjt||dd}||} |s[| |} | Sr )rQrrrr?rrrDrErrr7rrr.rre) r0rOr rr:r=shiftedrshifted_logsumexprr'r'r( _log_softmaxs(     rrRrw padding_idxscale_grad_by_freqsparsecCsV|dkrtd|d|jdkr'|d|}|jdkr%|d}|S||S)Nr z'weight' must be 2-D, got z-Drr)rOrr[ index_selectr)rrwrrrrr'r'r( embeddings     r num_weightsc Cstj|tjjd\}}||}t|tj}|r8||f}t |}t j ||g|dd}||} || d}t ||k|j} || d} ||f|j|jd} t j | |g| dd|S)NrTrrRr)rDrErrr7rr?longrLr;rrrTrVr[ masked_fillr) rirwr rrr:r=countsrqgrad_weights_scaleryr grad_weightr'r'r(embedding_dense_backwards&      r&cCsd}|D]}||9}q|SrZr')r0rIir'r'r(r#s rtensors num_chunksc Csg}|D]H}|}|||d||}|||kr7dgd|j|dd|||g}t||d}|d|t|dg}|||q|S)Nrrr rR)rr[rconstant_pad_ndr?Sizeappendr) r(rOr)padded_tensorstensor tensor_size pad_along_dimr view_sizer'r'r( _pad_chunk*s  r2cCs(|dj}|D] }|j|krdSqdS)NrFTr[)r(r[r.r'r'r(have_same_ndims?s  r4cCsB|dd|}|D]}t|d||kddq dS)NrcSr)NzG_chunk_cat expects same sizes of 0,...,dim-1 dimensions for all tensorsr'r'r'r'r(r}Lrz+leading_dimension_matches..)rr?r)r(rOleading_dim_sizesr.r'r'r(leading_dimension_matchesGsr6cCst|dkddtt|dkdd|dj}|dj}|D]$}t|dkddt|j|kddt|j|kddq"t|rVt|d |}nt|dkd d|D] }t||j kd dqbt |||S) NrcSr)Nz&_chunk_cat expects positive num_chunksr'r'r'r'r(r}Urz._preprocess_chunk_cat_inputs..rcSr)Nz0_chunk_cat expects a non-empty input tensor listr'r'r'r'r(r}WrcSr)Nz#_chunk_cat expects non-empty tensorr'r'r'r'r(r}\rcSr)Nz8_chunk_cat expects all input tensors with the same dtyper'r'r'r'r(r}_rcSr)Nz8_chunk_cat expects all inputs tensors on the same devicer'r'r'r'r(r}crcSr)NzK_chunk_cat expects non-negative dim when input tensors have different ndimsr'r'r'r'r(r}jrcSr)Nz3_chunk_cat expects dim < ndim for all input tensorsr'r'r'r'r(r}or) r?rr`rrlrr4rDrrOr[r6)r(rOr)expected_dtypeexpected_devicer.r'r'r(_preprocess_chunk_cat_inputsPs:   r9rcCsHt|||}t|||}|durt||dStj||d|d|S)Nr)r)r9r2r?r)r(rOr)rr-r'r'r( _chunk_catus  r: split_sizescCsXtj|||d}|durdd|DSt||D]\}}t||jt||ddqdS)Nr{cSsg|] }|jtjdqS)r8)ror?r<r/sr'r'r(r2sz)split_with_sizes_copy..Tr)rsplit_with_sizesrrrr)rwr;rOrsplitsrsplitr'r'r(split_with_sizes_copys  rA split_size.cCtj|||Sr5)rr@r)rrBrOr'r'r( unsafe_splitrDcCrCr5)rr>rd)rr;rOr'r'r(unsafe_split_with_sizessrFcs|j}||}dkr|dkrtd|d|fS|d}ddlm}||}fddt|D}|||d<t|||S) Nrz split_size is 0 but dim_size is z , expected 0r) guard_intcg|]}qSr'r'r/r'rBr'r(r2r~zsplit..rR)rrdetachrQrGrSr?r@)rwrBrOrHrfchunksrGr;r'rJr(r@s   r@tensor_indices_or_sectionscs|jjdkrtd|j|jtjkrtd|j|tdkp)dkfdddkrM|}t |t sGtdt|j | ||St }tj}r^|j}r^|j}|d d |D}Wdn1sswY| ||S) Ncpuz/tensor_indices_or_sections must be on CPU, got z.tensor_indices_or_sections must be int64, got rrcs ddS)Nz{tensor_split expected tensor_indices_or_sections to be a zero-dimensional or one-dimensional tensor, but got a tensor with z dimsr'r' split_dimr'r(r}szAtensor_split_tensor_indices_or_sections_py_impl..z%Expected sections to be IntLike, got cSsg|]}|qSr')itemrIr'r'r(r2zCtensor_split_tensor_indices_or_sections_py_impl..)rltyperrr?rrOrrQr-rr! tensor_splitr_guardsdetect_fake_mode shape_envignore_fresh_unbacked_symbols)rwrMrOsectionsctx fake_moderWrwr'rOr(/tensor_split_tensor_indices_or_sections_py_impls:        r\mat1mat2cCsH|s|st|}t|}|t||}|dkr|S|||Sro)r is_complexrhr?mm)rwr]r^rarjrr'r'r(addmms rause_gelucCs<t|||||}|r|jrtj|ddSt|St|S)Nr)r)rais_cudargelurelu)rwr]r^rarjrbrr'r'r(_addmm_activations    rfveccCs\|s|st|}t|}|t||}|dkr|S|dkr(||S|||Sro)rr_rhr?mvr)rwr]rgrarjrr'r'r(addmvs  rirrstdgammaNCHxWgroup output_maskc  stj|||ddtj||ddtj|ddt|kfddtjfkfddtdupJkfdd} t| kfddt||j dgd } |j dgd } d} d}d}| d r9d | }durt|  d  |  d}t|  d  |  d}t| d  d | }n&| |  d}| |  d}t| d tj d | f|j d}||||||}| |||}| d }t|d}t|d}t| | |t| | ||} | |j|j} | d r^| | | |  d | d j d gd  }| drj| j d gd }| ||fS)NF)allow_cpu_scalar_tensorscsddS)NzExpect input to have z elementsr'r')rmrnrlr'r(r}<rRz,native_group_norm_backward..csdddjS)NzExpect mean to have shape (, z , but got rr')rlrorr'r(r}@cs$dddurSdS)NzExpect gamma to have z elements but got rR)rr')rmrkr'r(r}D$csddS)NzExpect number of channels z, to be evenly-divisible by number of groups r'r')rmror'r(r}Jr~r r{rrdrRrrkr*)rDcheck_same_devicecheck_same_shaper?rrrrrtrrTrrqrlrVr7r)rirrrjrkrlrmrnrorpcpgdsdbd_inputd_gammad_biasr=ds_valdb_valc1c2c3r')rmrnrlrkrorr(native_group_norm_backward's        ""     $  rout2c  Cdt|||||||||| } | | | f}t| D]\}}|dur/t|||jt|||ddq|Sr)r enumeraterrr)rirrrjrkrlrmrnrorprrrrrWr'rIr'r'r(native_group_norm_backward_outs rcCs|dur ||S|Sr5r7)r0rr'r'r( _maybe_casts rgrad_outnormalized_shapebiasc"s|j}|} t|jfdd||||fD\} } } } | dur&td| t|}||d}|d|}g}g}t| D]}||krL||q@||q@t |}t |}ddl m }||dksl||dkr|dru| |nd|dr| ||dnd|dr| ||dfSdfSt || }t || }| durtd| ||}| dur| | }n| }||}t||d }t||}t||d }t||}|||}d}d} d}!|dr|||}|dr| durt|dkr t| ||d } n| |} |dr,| dur,t|dkr(t| |d }!n| }!t||jt| |dur;|jndt|!|durI|jfSdfS) Nc3s,|]}|dur|jtjdn|VqdSr)r7r?r<r.r9r'r(rs z-native_layer_norm_backward.. grad_out_cast should not be NonerrOrr zinput_cast should not be NoneTF)rrOrDget_computation_dtyperrr`rSr,rrQrPrLrVr?rrror)"rrrrrjrrrp input_shape input_ndim grad_out_cast input_cast weight_cast bias_castaxis inner_dims outer_dimsinner_dim_indicesouter_dim_indicesr'rlMrPx_hat grad_x_hatabrrrrKrzd_weightr|r'r9r(native_layer_norm_backwardsx                  rc Cs`t||||||||} || | f} t| D]\} }|dur-t| | |jt|| | ddq| Sr)rrrrr)rrrrrjrrrprrrrrWr'rIr'r'r(native_layer_norm_backward_outs rcCsg}tt|D] }|||dqt|j}||}|dur=|tj tj fvr5t tj j }n t tj j }n|}ttjjjt|dj|dd|} || } |durb| |} |jpk|duok|j} t|} | tjtjfv} | s| s| } | } | |}|| fS)Nrr Tr)rSr`r,rOrDrrr7r?rrfinforrrsqrtopsraddScalarrrr is_nestedr channels_lastchannels_last_3drr)rrrrdims_to_reducer'r:upcasted_inputeps_val rqrst_inputupcasted_resultrr9is_channels_lastrr'r'r(_fused_rms_norms:       rcCs|j}|}t|j}|j|tjd} |j|tjd} |dur)|j|tjdnd} | dur3td|t |} || d} |d| }g}g}t |D]}|| krY| |qM| |qMt | }t |}ddl m}||dksy||dkr|dr||nd|dr||| dfSdfSt|| }| dur| | }n| }d}d}| |}|drtj|||dd}|||||}|dr| dur| |}t |dkrtj||dd}n|}t||jt||jfS) Nr8rrrrTrF)rrOrDrrr7r?r<rr`rSr,rrQrrLrVrr)rrrrjrrprrr:rrrrrrrrr'rlrrrrzrrsum_vald_weight_full_shaper'r'r(_fused_rms_norm_backwardLsf             r running_mean running_varmomentum functionalc CsXdgttd|} t|j} |} |} |rt|j} |j| d} tj| | ddd\}}t ||}|||}t || }t || }|dur]||d||} |s]| | |dur| |j d}t || }|||d}||d||} |s| | nV|dus|durtd|j| dd}|} |j| dd}|} |}dt||}|jjd kr|}|}n |d }|d }t||d}t||d}|||}|dur|}t||d}||}|dur |}t||d}||}|jjd kr |j|jd}|j|jd}|j|jd||| | fS) Nrr rT)rO correctionrrz:running_mean and running_var must not be None in eval mode)rrrNr)rSrSrOrDrrr7r?var_meanrrcopy_rrrr@rlrSrLrVr)rrrrrrrrrreduction_dimsr:new_running_meannew_running_var input_acc biased_varrrjr save_mean save_rstdn squeezed_var unbiased_varinvstdr'r'r(native_batch_norm_helpersz                rr save_invstdc C,t||||||||d \}} } } } || | fSNFr rrrrrrrrrrrrUr'r'r(native_batch_norms  rc Csv|dur|durt||||||S|durtd|dur"td|r0t||||||||St|||||||S)Nz`running_mean is None, but running_var is provided. They should both be None or both be provided.z`running_var is None, but running_mean is provided. They should both be None or both be provided.)r_native_batch_norm_legitrR$_native_batch_norm_legit_no_training)rrrrrrrrr'r'r(native_batch_norm_decomposition s&  rcs||}||d|dkr6|dkr6fddt|D}||||d<tjjj|||Stjjj||S)NrrcrHr'r'r rJr'r(r24r~z(unsafe_chunk_py_impl..) rrSr?rrrFrdrDr)r.rLrOrfr;r'rJr(unsafe_chunk_py_impl.s rc Cstj|||||d||Sr)rrrd)rrrrrrrr'r'r(r:s rc Crrrrr'r'r(rPs  rc Cs,t|||dd|||d \}}}} } |||fSrr) rrrrrrrrrrUr'r'r(!_native_batch_norm_legit_no_statsas  rc CsPt||||||||d \}} } } } | durtd| dur!td|| | | | fS)NT#new_running_mean should not be None"new_running_var should not be None)rr) rrrrrrrrrrrrrr'r'r(#_native_batch_norm_legit_functionalpsrc CsPtj|||||d|}d}|tjjjkrtj||}tj|tj|j|j dS)a Return a reserve tensor for batch norm, used only by cudnn to pass forward state to the backward pass. This is needed for `_batch_norm_with_update` and `_batch_norm_no_update`, which support a variety of backends including cudnn. We create this tensor here to get the correct shape in the traced graph if we detect that will call the cudnn kernel, and rely on DCE to avoid materializing this tensor. Tr)rlayoutrl) r?_C_select_batch_norm_backend_BatchNormBackendCudnn(_get_cudnn_batch_norm_reserve_space_sizeemptyuint8rrl) rrrrrrrbackend reserve_sizer'r'r(_get_batch_norm_reserve_tensorsrc CsDt|||||d||d \}}} } } t||||||dd} ||| | fS)NTFrrr rrrrrrrrrrrUreserver'r'r(_batch_norm_with_update   rc Csht|||||d||d \}}} } } t||||||dd} | dur$td| dur,td||| | | | fS)NTrrr)rrr) rrrrrrrrrrnew_rmnew_rvrr'r'r("_batch_norm_with_update_functionals"rc CsDt|||||d||d \}}} } } t||||||dd} ||| | fS)NFrrrr'r'r(_batch_norm_no_updaterrcCsL|dur td|t||kjtjd}|||d|}||fS)Nz=generator must be None for _fused_dropout decomposition, got rrd)rr?r r7rr)rr4 generatorryr r'r'r(_fused_dropout_decompositionsr)rrrl pin_memory non_blockingr9rlrrr9c Cs |r|tjkrtd||rtdt|tjttttfs)tdt |j |durA|durA|durAt|tjr?| S|Sd}t|tjrL|}nt |}|durt||j krt|durl|j dkrltj||}d}tj|||}|dur|stj||}d}|durtj ||dS|S)Nz*layout must be None or torch.strided, got z:pin_memory=True is not supported in _to_copy decompositionz x must be Tensor or scalar, got FrNTr8)r?stridedrr-rrhrrrcomplexrSr!ro scalar_tensorrl_primsconvert_element_type device_put) r0rrrlrrr9dtype_convertedx_tensorr'r'r(_to_copy s6     rcCs t|Sr5)raliasr8r'r'r(nop_decomposition; s rout3exponential_average_factorepsilonc Cs^t||||||||\}} } |r|| | |jdtjdfS||d|d|jdtjdfS)Nrr)rrrLr?r) rrrrrrrrrrrr'r'r(cudnn_batch_normC s"   rcCs@t|D]\}}|dkr||jkr|j||ks||}q|SrZ)rr[rrT)r0broadcast_maskrryr'r'r(_broadcast_batch_norm_backwarde s   rrc Cst|||||||||| Sr5)native_batch_norm_backward) rrrrrrrrrrprr'r'r(batch_norm_backwardl src &s|j} |dur |j} n| } t|jfdd|||||||fD\} } }}}}}|j}|}|dkr.r z*rank of the input must be at least 2, got rz1mean and invstd must not be None in training modezDrunning_mean_cast and running_var_cast must not be None in eval moderd)rrDrrrOrrrSr?rrSr,rrrr7r)&rrrrrrrrrrpr weight_dtyperrrrunning_mean_castrunning_var_castsave_mean_castsave_invstd_castr input_rankr num_featuresrrrreduction_axesr'rgrad_output_sumdot_p grad_mean proj_scale grad_scaleprojrWr% grad_biasr'r9r(r s            rc  Crr)rrrrr)rrrrrrrrrrprrrrrWr'rIr'r'r(native_batch_norm_backward_out s& rsave_varc C t|||||||d|gd SNT)TTTrr)rrirrrrrrr'r'r(miopen_batch_norm_backward s r reserveSpacec Crrr) rrirrrrrrrr'r'r(cudnn_batch_norm_backward6 s rcs|j|jttdvfdd|jddD]}t|dkfddqd|ddkrjd|ddkrjtdd tdd|D}td d tdd||D}tjj |||Sd d d dfdd}|d|d\}}}} |d|d\} } } } |dt |d| f}| s| stj |ddSdd}||||| dd\}}||| | | dd\}} d}t t |jdt |jdD]\}}|dur|d|dd|f}q||d|dd|f}q||| S)Nrc dS)Nz9adaptive_avg_pool2d(): Expected 3D or 4D tensor, but got r'r'r3r'r(r}[  z%adaptive_avg_pool2d..r:rcdtdS)Nzjadaptive_avg_pool2d(): Expected input to have non-zero size for non-batch dimensions, but input has shape rrr'rr'r(r}` srRcss|] \}}||VqdSr5r')r/r'rr'r'r(rf z&adaptive_avg_pool2d..css&|]\}}}||d|VqdSrNr')r/r'rr=r'r'r(rg s cSstj|||ddS)Ntrunc rounding_moder?divrrrr'r'r( start_indexl sz(adaptive_avg_pool2d..start_indexcSs tj|d||d|ddS)Nrrrr r"r'r'r( end_indexo  z&adaptive_avg_pool2d..end_indexc stj|tjd}|||}||d}||}|dkp"||dk }|r+|d7}n|dkr3|d8}tj|tjd}|d|}|rbtj|d|j|jd} t|| }|||} | |} n|} || ||fS)NrmrrrRr)r?rprrTrrrlminimum) in_sizeout_sizeorangei0 maxlength in_size_modadaptive range_maxrxmaxvali1length)rlr$r#r'r( compute_idxr s(       z(adaptive_avg_pool2d..compute_idx.r*)rrRr{cSsjt|tr ||fS|dkrtd|||dk}|dkr$t|d}t||d}t|| }||fS)Nrz)dim should be negative when masking, got rRr:r*r)r-rrrTrVr?r")valsr1r.r-rOryr'r'r( maybe_mask s   z'adaptive_avg_pool2d..maybe_mask)r-rOr)rlrr`r?rrrnnr avg_pool2drVrr rS)rrrrTkernelr2idxhlength_h range_max_h adaptive_hidxwlength_w range_max_w adaptive_wr3r4retr'jr')rlr$r[rr#r(adaptive_avg_pool2dQ sN  (      & rBc Cstd|dt|jd| t|}tdd|Dr%||Sttj|jd| }ttj|}dg|j }|jd| |d| <|t j ||j d ||d}||} t j| d|g|ddd  | jS) N max_unpooling d_forward_outcss|]}|dkVqdSrr'r<r'r'r(r rz _max_unpoolnd..rrkrRFr)rDalert_not_deterministicrSranyrLroperatorrr[rrprlrtrr) rwrwrrO output_shapenchwindices_nc_shape indices_flatrr'r'r( _max_unpoolnd s"     rMcstjtjkfddttdkfddtjdvfddtjjkfddtdjD]td kfd dq>t dS) Nc djS)Nz2elements in indices should be type int64 but got: rr')rwr'r(r}  zmax_unpool2d..r cr)NzMThere should be exactly two elements (height, width) in output_size, but got elements.r`r'rr'r(r} rcdjdS)NzLInput to max_unpooling2d should be a 3d or 4d Tensor, but got a tensor with dimensions.r3r'rr'r(r} scdjdjSNzBExpected shape of indices to be same as that of the input tensor (z%) but got indices tensor with shape: rr')rwrwr'r(r}  rrcdjddS)NzZmax_unpooling2d(): Expected input to have non-zero size for non-batch dimensions, but got  with dimension being empty.rr')r'rwr'r(r} ) r?rrrr`r[rrSrrM)rwrwrr')r'rwrrwr( max_unpool2d s,          r]cstjtjkddtjdvfddttdkfddttdkfddttdkfddtjjkfd dtd jD]td kfd dqXtd d ko~d d ko~d d kfddt dS)NcSr)Nz(elements in indices should be type int64r'r'r'r'r(r} rzmax_unpool3d..r*rcrT)NzLInput to max_unpooling3d should be a 4d or 5d Tensor, but got a tensor with rUr3r'rr'r(r} rrcr)NzVThere should be exactly three elements (depth, height, width) in output_size, but got rPrQr'rRr'r(r} rScr)NzRThere should be exactly three elements (depth, height, width) in stride, but got: rPrQr'rTr'r(r} r~cr)NzSThere should be exactly three elements (depth, height, width) in padding, but got: rPrQr')rr'r(r}" r~crVrWrr')rwrr'r(r}& rXrrcrY)NzZmax_unpooling3d(): Expected input to have non-zero size for non-batch dimensions, but got rZr[rr')r'rr'r(r}/ r\r cr)Nz5strides should be greater than zero, but got stride: r'r'r`r'r(r}8 r) r?rrrr[r`rrSrrM)rrwrrTrr')r'rwrrrrTr( max_unpool3d sB             " ra)rjr.cCt||||d|dS)NTinplacerj _index_addr0rOr{r.rjr'r'r( index_add_> s rhcCrb)NFrcrergr'r'r( index_addJ s rirdc s"t|jtjdkfddjdkrdnd|jdkr*|ndtkfdddkr]t|jttupQt t fdd|}|jdk}|ri| dn|}df}|rwt j nt j} | |||dd } |r|S|r| dS| S) NrcrTNz(Index should have dimension 1 or 0 (got r&r3r'r{r'r(r}c rz_index_add..rcsdddS)NzNumber of indices (z') should be equal to tensor.size(dim) (z ), for dim=r'r')rO index_sizer/r'r(r}i csdtddS)Nzalpha argument of type z cannot be safely cast to type !)rSr')rj python_typer'r(r}p rmr5Tr)rDcanonicalize_dimsr[r?rr dtype_to_typerrris_weakly_lesser_typerSrTr index_put_ index_putrr) r0rOr{r.rdrjzero_dimr6rxrtrr')rjrOr{rlror/r(rfW s6     rfrc Cstt|dkddt|}|d}|dd}tdd|D}|r,||f}n||f}||}|d||}dt|} t|D]+} || } t| | d|| df|} |rhtj || d| d}qFtj || d| d}qF|S) NrcSr)Nz#received an empty list of sequencesr'r'r'r'r(r} rzpad_sequence..rcss|]}|dVqdSrrr.r'r'r(r zpad_sequence..)rrrOr{) r?rr`rrrrSrr*r|) sequences batch_first padding_valuesequences_sizemax_size trailing_dimsmax_lenout_dimsr dim_paddingsr'currseqrowr'r'r( pad_sequence s(     rcCt||||ddS)NTrd _index_copyr0rOr{r.r'r'r( index_copy_ rcCr)NFrrrr'r'r( index_copy rrc st|j|}tjdkfdd|jdk}|r |dn|}jdkr,dnd|f}|r:tjntj}||||} |rG|S|rN| dS| S)NrcrTrjr3r'rkr'r(r} rz_index_copy..rr5) rDrpr[r?rrTrrsrtrr) r0rOr{r.rdrur6rxrtrr'rkr(r s   rcCsRt|d|}tt| }|js|jr|d}n|}|t||fS)Nr'r)r?r&rLrerrcis_xpur-)rwrrgrr'r'r(log_sigmoid_forward s   rlowhighrc Cs*tj|j|t|t||j|j|dS)N)rTrrrrlr)prims_uniform_helperrrTr rrl)r0rrrr'r'r(uniform srcCs|t||||Sr5)rr)rwrrrr'r'r(uniform_ srcCst|d}|dur"t|duddtt||kdd|S|durjt|duddtt||kddg}t|D]%\}}t||krZ|||dt|qB|t||d|qB|StddddS) Nr cSrNz9Must specify exactly one of output_size and scale_factorsr'r'r'r'r(r} rz.upsample_compute_output_size..cSrNr'r'r'r'r(r} rcSrrr'r'r'r'r(r} rcSrrr'r'r'r'r(r} rFcSrrr'r'r'r'r(r} r)r`r?rrrhr,r )rr scale_factorsspatial_dimensionsr'r=r'r'r(upsample_compute_output_size s.  rcCs|durdS||Sr5r')scalesrxr'r'r(get_scale_value srrcCs2t|||}|r |ndgt|}t|||Sr5rrr`_upsample_nearestrrrosizerr'r'r(_upsample_nearest_vec s rcCs6t|||}|r |ndgt|}t|||ddSNTexactrrr'r'r(_upsample_nearest_exact_vec srcCsg}t|}|r dnd}t|D]O}||}|j| |} ||dur2||dkr2| | ||n| |} tj|tj|jd} | || tj} t|d|D]} | d} qR| | q|S)NrrrrrrR) r`rSrr?rprrlr7rrTr,)rrrrrwnum_spatial_dimsr~rrisizerkoutput_indices input_indicesrUr'r'r(!_compute_upsample_nearest_indices- s      r)preserve_memory_formatrrcCst|||gSr5rrrrr'r'r(upsample_nearest1dS s rcCst|||gddSrrrr'r'r(upsample_nearest_exact1d_  rscales_hscales_wcCst||||gSr5rrrrrr'r'r(upsample_nearest2dm s rcCst||||gddSrrrr'r'r(_upsample_nearest_exact2dz s rscales_dcCst|||||gSr5rrrrrrr'r'r(upsample_nearest3d rrcCst|||||gddSrrrr'r'r(_upsample_nearest_exact3d s rrc Cspt||||d}ddg|}t||}|jdkr6t|}|jd}|jjdkr0|dkr0t j }|j |d}|S)Nrr*rcudar8) rr _unsafe_indexr[rDrrrlrSr?r<r) rrrrspatial_indicesrwrr9 n_channelsr'r'r(r s      rcsn|r|rdn |r dn|rdndtdkr'tdtdfdd tdtDS) Nrr*rr rz len(params)=z is not divisible by group_size=cs g|] }t||qSr'rrI group_sizeparamsr'r(r2 sz!gather_params..)r`rrS)r has_biaseshas_projectionsr'rr( gather_params s rcCsh|r!|d||d|}}|d|d|d|d}}n ||||}}d\}}||||fS)Nr rNNr')rhiddensr' bidirectional cur_params cur_hidden bidir_params bidir_hiddenr'r'r(params_hiddens s $ rcCsD||krtd|d|d||d||||dd|S)Nlast_batch_size (z) must be > batch_size (r&r)rr,r)rlast_batch_size batch_sizerr'r'r(update_hidden_for_packed s rc CsF||kr|S||krtd|d|dt||d|||fS)Nrz) must be < batch_size (r&r)rr?concatr)rrr inp_hiddenr'r'r( update_hidden_for_packed_reverse srcCs$|d}|d}|r|dnd} |r|dnd} g} g} |r"|dn|d} |dd| }t|t|}|r>|ddd}|D]-}|jd}| |krLn|rVt|| ||}nt|| || }||||| || }|} | |q@|ru| n | || t | d}|st | dn|}||fS)Nrrr rrR) rr?r@rSrrrr,reverser)inphiddenrr hidden_fn batch_sizesr ih_weight hh_weightih_biashh_bias step_outputrrr split_inpr'r hidden_outr'r'r(one_layer_rnn_data s@     rcfdd}|S)Ncst||||Sr5rlinearr'rrrrr nonlinearityr'r(rK+ szrnn_cell..innerr'rrKr'rr(rnn_cell* s rcr)Ncs$t|||}t||||Sr5rrrr'r(rK2 szrnn_cell_data..innerr'rr'rr( rnn_cell_data1 s rcCs|d}|d}|r|dnd}|r|dnd} t|||} |r&| dn| } |d} g} | D]} || | |||| } | | q1|rH| t| d}|| dfS)Nrrr r) rrfliprTr,rr?rr)rrrrrrrrrrprecomputed_inputrrr'rr'r'r( one_layer_rnn9 s   rcCs|d}|d}|r|d}|d}nt|}t|}|dd} |dd} g} d} | d} d}d}d}d}|}| } | } tjjj|||||| | || | | |||||}|d|d|d}}}|| d| dffS)Nrrr rF) r?rrrTrrrmkldnn_rnn_layerrdr)rrrrrw0w1w2w3hxcxrmode hidden_size num_layersrrzroutputsrYhycyr'r'r(mkldnn_one_layer_lstmO sN   rc Cs|r|ddn|}g} t|D]^} t||| |\} } }}|r'|| dkr'|nd}| || | |\}}| ||rI| ||||dd\}}| ||rXt||g|d}n|}|dkrn|rn| |dkrntj||dd}q|rw|ddn|}|| fS)NrrrT)r)r) transposerSrr,r?rrOr)rrrrrrrrrzlayer_fn final_hiddensr'rrrrfwd_inp fwd_hiddenbwd_inp bwd_hiddenr'r'r( _rnn_helper s,      rc CR|d} t||d}t|| |||||||ttttjd \} } | t| dfSNrFr) unbindrrrrrr?rstack rrrrrrrrrzrrrr'r'r(rnn_tanh_input    rc Crr) rrrrrrr?rerrr'r'r(rnn_relu_input r r c CT|d} t||d}t|| ||||||dtt|ttjd \} } | t| dfSNrFrr) rrrrrrr?rer datarrrrrrrrrrrr'r'r( rnn_relu_data &  rc Cr r ) rrrrrrr?rrrr'r'r( rnn_tanh_data rrcCst||||}|d|}|d} |d} |d} |d} | || | } | | }|dur;|nt||d}|| fS)Nr*rrr rrrchunkrr)rrrrr hr_weight chunk_dimgates chunked_gatesin_gate forget_gate cell_gateout_gaterrr'r'r( lstm_cell.s      rc Cs|d}|d}|r|dnd}|r|dnd}t|dkr"|dn t|dkr,|dnd} |dd} |dd} t|||} |rJ| dn| } g} | D]}t|| | ||| dd\} } | | qP|rk| t | d}|| d| dffS)Nrrr rrr*r) r`rTrrrrr,rr?rr)rrrrrrrrrrrrrrrr'r'r(one_layer_lstm<s$*  rc Cs |d}|d}|r|dnd}|r|dnd} t|dkr"|dn t|dkr,|dnd} g} g} |r8|dn|d} t|t|}|rM|ddd}|d}|d}|dd| |dd| }}|D]l}|jd}t|||}|| kr| |d|| ||d|| |f|dd||dd|}}|| krt ||d| || fd}t ||d| || fd}t ||||| | dd\}}|} | |qf|r| ||f}n| ||f| t | \}}t |dt |df}t | d}||fS) Nrrr rrr*rRr)r`r?r@rSrrrrr,rrrrr)rrrrrrrrrrrrrrrorig_hxorig_cxrrr'rhidden0hidden1rr'r'r(one_layer_lstm_dataWs\*       r$cCsdd}||||r tStS)a*Check whether we could use decompose lstm with mkldnn_rnn_layer. All the below conditions need to be met: * ``torch._C._get_mkldnn_enabled()`` returns ``True``. * All the input args are on CPU. * The dtypes of args are either torch.float or torch.bfloat16. * Inference. * ``has_projections`` returns ``False``. Args: * input: the input sequence to LSTM * hx: a tuple of the input hidden state and cell state ``(h_0, c_0)`` to LSTM * params: the weight and bias tensors of LSTM c SstjsdS|gt|tt|}dd|D}t|dkr$dS|}|tdkr1dSdd|D}|D] }|tj tj fvrGdSq:|j rMdS|d d|d dk}|r_dSd S) NFcSh|]}|jqSr'rkr/tr'r'r( zEselect_one_layer_lstm_function..use_mkldnn..rrNcSr%r'rr&r'r'r(r(r)rr T) r?r_get_mkldnn_enabledrSr from_iterabler`poprlrbfloat16 requires_gradr) rrrr(devicesrldtypesrrr'r'r( use_mkldnns(  z2select_one_layer_lstm_function..use_mkldnn)rr)rrrr1r'r'r(select_one_layer_lstm_functions r2c Cst|dkrtdt|t|||dd|ddk}tt|d|d} t|||} t|| |||||||| \} } tt| } | t | ddt | ddfS)Nr $lstm expects two hidden states, got rr) r`rrrrSrr2rr?r) rrrrrrrrrzrrrrr'r'r( lstm_impls& $  "r4c Cst|dkrtdt|t|||dd|ddk}tt|d|d} t|| ||||||dtt|d \} } tt| } | t | ddt | ddfS)Nr r3rrF)r) r`rrrrSrrrr$r?rrr'r'r(lstm_data_impls$ $  "r5c Csr|dd}t|||dd}|d|d}|d|d} |d|d|} || | | S)Nrrr r)rrrrr rrrrrrchunked_igateschunked_hgates reset_gate input_gatenew_gater'r'r(gru_cells r<c Cs|t|||dd}t|||dd}|d|d}|d|d} |d|d|} || | | S)Nrrrr rr6r'r'r( gru_cell_datas r=c CsJt||d}t||d||||||dtt|td \} } | t| dfS)NFrr )rrrrrr=r?r) rrrrrrrrrrrr'r'r( gru_impl_data s   r>c CsHt||d}t||d|||||||tttd \} } | t| dfS)NFrr)rrrrrr<r?r) rrrrrrrrrzrrr'r'r(gru_impl>s   r?cC:t|||}t|d}t|d}tjj|||||SNrr)rrrr?rr_upsample_bilinear2d_aarr align_cornersrrscale_hscale_wr'r'r(upsample_bilinear2d_aa_vec\    rGcCr@rA)rrrr?rr_upsample_bicubic2d_aarCr'r'r(upsample_bicubic2d_aa_vechrHrJcCr@rA)rrrr?rr_upsample_lanczos2d_aarCr'r'r(upsample_lanczos2d_aa_vectrHrLcCs4t|||}|r |ndgt|}t||||Sr5)rrr`_upsample_linear)rrrDrrrr'r'r(_upsample_linear_vecs rNrDcCst||||gSr5rM)rrrDrr'r'r(upsample_linear1dsrPcCst|||||gSr5rO)rrrDrrr'r'r(upsample_bilinear2ds rQcCst||||||gSr5rO)rrrDrrrr'r'r(upsample_trilinear3ds rRcCs@|r|dkr|d|dSdS|dur|dkrd|S||S)Nrrdrr')r'r(rDrkr'r'r(_compute_scales rScCs|r||S||ddSNrr')rk dst_indexrDr'r'r(_compute_source_indexsrVweightsweights_precisioncCsBtddt||Dd|d>}||?}t|ddtjS)Ncss,|]\}}|tj|tjVqdSr5)r7r?r)r/r=rr'r'r(rs z%_sum_tensors_uint8..rr) _sum_tensorsrr?rr7r)rjrWrXrr'r'r(_sum_tensors_uint8s r[cCsJt|}d}tj||jd}d|d|d>}|dk}||S)Nrkrri)r?rrrprlr)rW max_weightmax_weight_precision precisionsvaluesryr'r'r(_compute_weight_precisions  racsjd}jdd}t|tjtjjd\}fddfddtt|||D}tt|\}g} t ddggD]#ddgfd dt D} t | } t | } | | qGtt D]'} || | d d fd dt| ddd| dddD} qqt| dkrtd t| | d} t}jjdkr|dkrtj}t| tjstdt| j| j|d} s| } | S)Nrr rc st|||}tj|jdjd}t||jdd}|j|jdgdg|R}|tj }|dj|dd}|||fS)Nrkrrrrrr) rSr?rprlr7rVrrrr) inp_sizer(rnsqueeze scale_factorr'x_f32r0xp1)rDrrr'r( get_valuess  z$_upsample_linear..get_valuescs,g|]\}\}}}|||d|qSr r')r/r'rbr(r)rgn_dimsr'r(r2s z$_upsample_linear..rcs(g|]}|dkr|n|qSrr')r/k)rxp1sxsr'r(r2(rrdcs$g|]\}}|t||qSr')r?r)r/v1v2)xscaler'r(r2 sz+Expected vs to have exactly 1 element, got rz$Expected result to be a Tensor, got r8)rr`rDrEr INT_TO_FLOATrrrSr rSrrrr,reversedrr7rrrlrSr?r<r-rr!rrround)rrrDrr inp_sizesrUr`xs_f32vsrxvr'rr9r') rrDrrgrrhrjrkror(rMsN    "        rMrrcCs |j|jkSr5r)rrr'r'r( is_same_size(rzrxcGrtr5)rrt)r0rrFr'r'r(_reshape_alias-s rycCrtr5)rr{)r0rwr'r'r(r3rzrcCst||||Sr5)rrt)r0rwrxrr'r'r(r8rErcCs|D]}|durt|jtjtjfvddqt|jtjkddddlm}||dkr@tj ||}| |j |St t|D]}||}|dur^|jd||dd||<qFt||||S)NcSrNz3tensors used as indices must be long or int tensorsr'r'r'r'r(r}Crz&_unsafe_masked_index..cSrNz*tensors used as masks must be bool tensorsr'r'r'r'r(r}Hrrrrrb)r?rrr!rhrrrQrr_meta_registrationsmeta_index_TensorrrrSr`rrrrr")r0ryrwfillr{r meta_resultr'r'r'r(ru=s*  rucCs|D]}|durt|jtjtjfvddqt|jtjkdd|dkr.|Stt |D]}||}|durP|j | | | |dd||<q4| |d}t j|||ddS) NcSrrzr'r'r'r'r(r}_rz5_unsafe_masked_index_put_accumulate..cSrr{r'r'r'r'r(r}drrrrbTr)r?rrr!rhrrrrorSr`rrr"rr)r0ryrwr`r{r' masked_valuer'r'r(#_unsafe_masked_index_put_accumulateYs(  $rcCsV|}d}|dkr d}|dur,|dkr&dg|}|jd||<||}n|}||}t||k|d} | |} t||| | } t||k| d} |tj j krb|dkrb| dd} | | fS|dur| |j}t||| |} t||k| d} | } n ||k |} |tjj kr| } | | fS|tjj kr| | } | | fS)Nrr rr'r)rOrrtr?rfrTgatherrrr$rxrrnrr7r&r%)rwrrrr rhrrwr safe_target_rr wsumr'r'r(_nll_loss_forwardssB        rcsdkr dkstdddkr&tdddko1dk}|sGtjdjdkfddjd }|durg|dkr\||ksgtd |d |jt|||S) Nrr r!r"rr#csdjdjdS)Nr$r%r&rr'rwrr'r(r}rmz"nll_loss_forward..rRz/weight tensor should be defined either for all z7 classes or no classes but got weight tensor of shape: )rOrr?rrrr)rwrrrr r( n_classesr'rr(nll_loss_forwards(     rcCst|||||Sr5)r)rwrrrr r'r'r(nll_loss2d_forwards rAcCs |d||d||dS)Nr rrr'r0rr'r'r(_upsample_cubic_convolution1r%rcCs(||d||d||d|S)Nrr*r'rr'r'r(_upsample_cubic_convolution2s(rr'c Csd}|jtdkrDtj|d|gdd}tj|dd|gdd}t||}t||}tj|dd\}}tj|dd\}} ||| |fSt|d|t||td||td||fS)NgrNrdrr{r)rlr?rrrr) r'rtt1tt2w03w12rrrrr'r'r( _upsample_get_cubic_coefficientss      rcoeffstscCs t|}tddt||DS)Ncs|] \}}||VqdSr5r'r/rrr'r'r(rrz+_upsample_cubic_interp1d..)rrZr)rrcoeffs2r'r'r(_upsample_cubic_interp1dsrcCs ttj|Sr5)rr?r)rr'r'r(rZs rZ num_stepscCsB|dkr tjd||dS|s|d|nd}tj| ||||dS)Nrrrm)stepsrlr)r?r.linspace)rrDrrlrr'r'r(_linspace_from_neg_onesrthetahrc Cs|j}|j}t||||d|d}t|||||dd}tjd||d}tjjj|dddd}tjjj|dddd}tjjj|d ddd}|||S) Nr)rrrr)rr constantrrrrxrr)r r rrlrrtr?rqr5rr) rrrrDrrlgrid_xgrid_ygrid_oner'r'r(_make_base_grid_4d s rrc Cs|j}|j}t||||dd|d}t||||d|dd}t|||||ddd} tjd||d} tjjj|dddd}tjjj|dddd}tjjj| d ddd} tjjj| d ddd} ||| | S) Nr)rrrrr)rrrrrrr r)rrr) rrrrrDrrlrrgrid_zrr'r'r(_make_base_grid_5dsrc CsL|\}}}}t||||d}|ddd|jdd}||||dS)NrDrRrrr:r )rrtr>rTr) rrrDrrUrr base_gridgridr'r'r(_affine_grid_generator_4d+s  rc CsR|\}}}}}t|||||d}|ddd|jdd} | ||||dS)NrrRr*rr:r)rrtr>rTr) rrrDrrUrrrrrr'r'r(_affine_grid_generator_5d5s rcCs@tt|dvddt|dkrt|||dSt|||dS)Nr^cSr)NzCaffine_grid_generator needs 4d (spatial) or 5d (volumetric) inputs.r'r'r'r'r(r}Erz'affine_grid_generator..r*r)r?rr`rr)rrrDr'r'r(affine_grid_generator?s  rrinterpolation_mode padding_mode _expand_gridcsTtdvfddtdvfdddtdtdtffdd dtd td tdtfd d dtdtdtffdd dtdtdtffdd }j\ |j\}}|dkrhtd|rz|d|d}dtdtdtf fdd tjj ddddtjj dddddtdtdtdt ffdd dtdtdtffdd |d!} |d"} d#kr6|| } || } | | d} }d}}| |}}|| || }| ||| }| | | |}| | }t fd$d%|f| ||f|||f|||ffDSdkrS|| } || } | }| } ||dS| } | } | | | | }szd|d}dtdtdtf fd&d' d(tdtf fd)d* t fd+d%td,D}t||S)-N)rrr cr)NzInvalid interpolation mode r'r')rr'r(r}^rz"_grid_sampler_2d..cr)NzInvalid padding mode r'r')rr'r(r}arcoordsrrPcs0r|ddn|d}|dd}|||SrTr')rrrofsrr'r( unnormalizeds  z%_grid_sampler_2d..unnormalize twice_low twice_highcSsv||kr t|S|d}||d}||}t||}||jtjd}t|d@dk|||||S)Nr rrr)r?rrfmodfloorr7int8rf)rrr coords_min coords_spancoords2extraflipsr'r'r(reflect_coordinatesos    z-_grid_sampler_2d..reflect_coordinatescsfdkr|Sdkrt|d|dSr |dd|d}n |dd|d}t|d|dS)Nrrr rRr)rrcoords_reflected)rDrrr'r(compute_coordinates{sz-_grid_sampler_2d..compute_coordinatescs||}||Sr5r')rr coords_un)rrr'r(compute_source_indexs  z._grid_sampler_2d..compute_source_indexr z4grid last dimension must be 2 (for x,y coords), got rrkysc s,td|kt|ktd|k|kSror?rs)rkr)iHiWr'r(in_bounds_conds$z(_grid_sampler_2d..in_bounds_condrkwscsN||r ndtfdd|jtjd|jtjd|fDS)Nrc3s*|]}t|dVqdSr)r?rfrtr&)rlrroHoWr'r(rs  z1_grid_sampler_2d..clip..r)rr7r?r)rkrr)rmrlrrrr)rrr(clips  z_grid_sampler_2d..clipixiycs&|||\}}}||f|Sr5r')rrridx_xidx_yw_)C_idxN_idxrrr'r( get_summandsz%_grid_sampler_2d..get_summand).r).rrc3s"|] \}}}|||VqdSr5r')r/rrr)rr'r(rs  z#_grid_sampler_2d..cs |}|}||dSrZr')rrr0rY)rrrrr'r(get_value_boundeds   z+_grid_sampler_2d..get_value_boundedrcsF|d}d||d|d|f}t|S)Nrr )r)riy_ofscs)rix_nwiy_nwtxr'r( get_coeffs     z#_grid_sampler_2d..get_coeffc3|]}|VqdSr5r')r/r)rr'r(rrr*)r?rrrhrrrtrnrprlrrrZrsrTrrSr)rrrrrDrrrUtwor0rYrrix_neiy_neix_swiy_swix_seiy_sew_nww_new_sww_se ix_nearest iy_nearesttyrr')rmrrlrrrrDrrrrrrrrrrrrrrrrrr(_grid_sampler_2dMs~     (               rcCst|||||dS)N)rrrrD)r)rrrrrDr'r'r(grid_sampler_2ds rcs`tdko dkfddtddkfddjddS)Nr rcsddS)Nzmatrix @ vector expected, got rrr{r'rwrgr'r(r}rszmv..rcs*dddddddS)Nzsize mismatch, got input (rr0rz), vec (r&rvr'rr'r(r}s*r{)r?rrOrrrr'rr(rh s  rhcCsp|dur|d|d}d|||t|}n d||t|}|dur-||}||j}t||SrZ)r logsigmoidr7rr)rwrr pos_weightr log_weightrr'r'r( binary_cross_entropy_with_logitss  rtensor1tensor2is_outc s|j|jkr ||fn||f\}}ddlm|jdkr |jdks"dS|jr)|s)dS|jdkr0dS|dkr:dS|j}|}dg}t|ddD] }|||dqLt fd d t |t t||DS) Nrrrr FTrrRc3s.|]\}}}|dkp||kVqdSrr')r/rrrrr'r(rDs  zshould_fold..) r[rQrr.rrrTrrr,rrrS) rrrt1t2t1_shape t1_strideexpected_striderr'rr( should_fold,s(     r) pass_is_out)rcCsddlm}m}|}|}|dks|dkr"td|d||dkr0|dkr0t||S|dkr>|dkr>t||S|dkrT|dkrTtt t |d|dS|dkrb|dkrbt ||St |||r||k}|rq|j n|}|sw|n |dkr| n|} |j} t| dd} ttj| } | dk} | r| | jd|| | d}| rtjj| | | }|r|j S|Stjj|| | S|dkr|dkr|dkr|dnd}|d}|jdd}|dkr|dn|d}|dkr|dnd}g}t|dD] }|||q|d krY|d krY||d|dkrY||ddkrD|jrDt|d|S||ddkrY|jrYt||dStt||}|||g}t|}|||||}|dk}|r||g}|||| d}n|||g}|||||}|} |dkr| ||dkr| ||r| |d!| S| |!| St"d d d dS) Nr)r guard_or_truez9matmul does not support 0-dimensional tensors, got dims: z and rr rRr:rFcSr)Nz/both arguments to matmul need to be at least 1Dr'r'r'r'r(r}rzmatmul..)#rQrrrOrr?dotrhrr`rTrr>r'rrSrrGrr,rrr _unsafe_viewrrrSr.r=broadcast_shapesrrnbmmrtr)rrrrr dim_tensor1 dim_tensor2rrrsizes_1rH folded_dim1 t2_is_matrix t1_foldedrrm1 batch_tensor1m2r4 batch_tensor2r'expand_batch_portiontensor1_expand_sizeexpand_batch_producttensor1_expanded vector_rhstensor2_expand_sizetensor2_expandedr'r'r(r=Ls                    r=rErFcsj\}}t|d||}t|d||}tjtjjd\}}tj|djdj |d} tj|djdj |d} t || |} t || |} | d} | } | }| | dd}| |  dd}| tj} | tj}|d||d|d f}| d| | d| d ft|t|}d \jtjkrtt|fd d Dfd d |D}fddfddtfdd|D}jtjkrdurtdt||}n tddt||D}t}|j|d}|S)NrrrrkrrRrrdr rc.g|]}|d>t|dtjqSrrr?rr7int16r/r)weights_precision_xr'r(r2 z.upsample_bicubic2d_default..crrrr)weights_precision_yr'r(r2rcs<t|dd}t|dd}tdd||g}|SrA)r?rrr)rrky_idxx_idxrw)in_hin_wrr'r( load_boundedsz0upsample_bicubic2d_default..load_boundedcsXtfddD}jtjkr durtdt|Stddt|DS)Nc3s|]}|VqdSr5r')r/x_ofs)r!rYr'r(rrwzCupsample_bicubic2d_default..get_x_interp..z4weights_precision_x must not be None for uint8 inputcsrr5r'rr'r'r(rr)rrr?rrr[rZr)rYsrc_x)rixs_ofsr!r weights_x)rYr( get_x_interps  z0upsample_bicubic2d_default..get_x_interpc3rr5r')r/y_ofs)r&r'r(rrz-upsample_bicubic2d_default..z4weights_precision_y must not be None for uint8 inputcsrr5r'rr'r'r(rrr8)rrSrDrErrqr?rprlr7rVrTrrrrrrrarrr[rZrrr)rrrDrErFrUh_scale_factorw_scale_factorrr'rAx_floaty_floatr0rYyscaleroiys_ofs weights_ysrc_yrr9r') r&rr rr$r!rrr%r(upsample_bicubic2d_defaultsT             r0cCstt|t|dkdd|dur4|durtdttttftddt|jdd|D}|r8|nd\}}t |||||S) NrcSr)Nz:Must specify exactly one of output_size and scale_factors.r'r'r'r'r(r}3rz(upsample_bicubic2d_vec..z7scale_factors must not be None when output_size is Nonecss$|] \}}tt||VqdSr5)r r )r/rrkr'r'r(r<s  z)upsample_bicubic2d_vec..r r) r?rrrrr rrhrrr0)rrrDrrErFr'r'r(upsample_bicubic2d_vec&s"  r1cfdd}t||S)Ncs4tj| ||jd}|d|d|S)Nrkr)r?rprlrrmiddlerdim_idxrr'r(rxKsz_reflection_pad..idx_reflection_or_replication_padrrrxr'r6r(_reflection_padE r:cr2)Ncs*tj| ||jd}t|d|dS)Nrkrr)r?rprlrr3r6r'r(rx\sz_replication_pad..idxr7r9r'r6r(_replication_padVr;r<idx_fnc stdt|ddfvfdd|j d}|}fddtD}fddtD}|}tD]}dg|} |||||||| ||<t|| }qFt |} |j | d}|S) Nr rcs ddddddS)Nreflection_padz d requires rzD or r zD inputr'r'r{r'r(r}o z0_reflection_or_replication_pad..c g|] }dd|qSrr'rIrOrr'r(r2tr?z2_reflection_or_replication_pad..c$g|]}dd|dqSrr'rIrAr'r(r2urtr8) r`r?rrOrrSrrrDrr) rrr= inp_shapenc_dim padding_left padding_rightrr'rxr9r'rAr(r8gs"       r8cs\tddd|j dDfddtDfddtDg}t|jD]}dg|j}d||<|tj|j||jd|q2|d | d d d  fd dtD fd dtD} fd dtD}fddtD t t j  fddtD}t |d} fdd} tjddtDD]f} | tdgkrqg} g} tD]K}| |dkr|} |}n0| |dkr||} |d|f}n| |dkr||} ||||df}| || |q| | | | } q| S)Nr cSsg|]}|dqSr r')r/rr'r'r(r2rRz,_reflection_pad_backward..cr@rr'rIrAr'r(r2r?crBrr'rIrAr'r(r2rtrrRrkcSs|\}}}t||k||kSr5r) index_ranger'lbubr'r'r(index_range_conditions z7_reflection_pad_backward..index_range_conditioncsg|] }||qSr'r'rIrExyzr'r(r2rcsg|] }||qSr'r'rIrKr'r(r2rcs(g|]}d|||qSrr'rI)dhwrErLr'r(r2rlcs.g|]}|d|||fqSrr'rI)centerrMrErFr'r(r2s"csg|]}|qSr'r'rI)rJrange_cr'r(r2rsrcsttD]}||d||dk}t|tr|r|Sqttjfdd|D}t||d}||S)Nr rcsg|]}|qSr'r')r/rG)rJr'r(r2rRz@_reflection_pad_backward..accumulate..r)rSr-rrrLrrrsru)rr index_rangesr'upper_less_than_lowerrg)rrOrirJr'r(rs z,_reflection_pad_backward..accumulatecSsg|]}gdqS))rRrrr'r r'r'r(r2rRr)r`rrSr[r,r?rprlrtrLrrrsru itertoolsr r)rir0rrwr' view_shape left_reflect right_reflectrrrareaoutsrPrrGr') rrNrMrOrirJrrErFrOrLr(_reflection_pad_backwardsT  $    "  rYrrrcCs(tj|||d}tj|||d}||fS)Nr)r?aminr)rwrOrrZrr'r'r(aminmaxsr[rcCs"tjtt|d||||dS)Nrr)rrr?rfisnan)rwrOrrr'r'r(nansums"r]rrrlrrc Cstjjd|d||||dS)Nrrr^rrp start_step)rJrrrlrr'r'r(arange_default rac Cstjj||d||||dS)Nrr^r_)rIrJrrrlrr'r'r( arange_startrbrccOsddlm}||i|S)Nr)out_dtype_dense)!torch._higher_order_ops.out_dtyperd)rFrGrdr'r'r(out_dtype_decomp s rfmarginc sttjdjdt|dkp|dkddtjdko,dkfddtjdko?kfdddurdttjdko\kfdddtjdd }||}| d}|dkr|n||}dur|}tj j d }t |k|d}|t jjkr|S|t jjkr||jdS|jdd S) Nrrr cSr)Nz only p == 1 and p == 2 supportedr'r'r'r'r(r}"rz#multi_margin_loss..crNNzMExpected non-empty vector or matrix with optional 0-dim batch size, but got: rr'r_r'r(r}%rOcddjS)Nz#inconsistent target size, expected rrr')nframerr'r(r})r)cri)Nz#inconsistent weight size, expected rrr')rOrr'r(r}/r)rxrkr{)r? atleast_2d atleast_1drrr[rrTrr?rprlrfrr%rxrr&r) rrr4rgrrurgrxr')rOrrjrrr(multi_margin_losssB              rn is_targetc s|j|jt|}t|}|jd}ttdko |dkfddttdko2kfddtj||jd}|dk}tjt|||dd d }||k}t||d}tj |d|d } t||d} tj || j dd kdd } d | j j dd |} | d} | |} t| d| } |tjjkr| jdd } n|tjjkr| } n| jdd } | |j} | | fS)Nrr rcrrhr'r')orig_input_shaper'r(r}Qrz0multilabel_margin_loss_forward..csddS)Nzinconsistent target size: z for input of size: r'r'rporig_target_shaper'r(r}Ur~rkrRTrrxr{rd)rrR)rr?rkrr`rprlrZrfrrFrTTr?rr%rxrrr&r7rr) rrrrOrxis_endend_idx target_masktidx0rmtidx1rorgr'rqr(multilabel_margin_loss_forwardBs@          ry) attn_maskrkquerykey dropout_p is_causalrzc sttfddtdko dko dkfddtdkfddtjdjdkoJjdjdkddtjj||d|d d kd \}}| d d d dj tj d  d d d d}||fS)NcrN)Nz-query must be FP32, FP64, BF16, FP16 but got rr')r{r'r(r}rOz.r*cs"dddS)Nz,q, k, v must be a 4 dimensional tensor, got rrr{r')r|r{rxr'r(r}s"rcr)Nz&dropout probability must be zero, got r'r')r}r'r(r}rrcSr)Nz&q, k, v should have the same head sizer'r'r'r'r(r}rr)rzr}r~ dropout_maskrk enable_gqar rr8) r?rrrOrr"_scaled_dot_product_attention_mathrdrrrr<) r{r|rxr}r~rzrkrattnr')r}r|r{rxr(*scaled_dot_product_flash_attention_for_cpus@  "& # rcst|fdd}|S)Ncs|i|}|d|Sro)r)rFrGr outplace_opr'r( inplace_opsz$register_inplace..inplace_opr)aten_oprrr'rr(register_inplacesrcCsx|s|st|}t|}t||}t|tjr |dkr$||}|dkr*|St|tjr4|dkr8||}||S)Nrr)rr_rhr?rr-numbersNumber)rwbatch1batch2rarjrr'r'r(baddbmms rcCstj||ddS)Nrrr )rwrr'r'r( floor_dividesrcCsttj|jdSrZ)rLrrGrr)r'r'r'r( sym_numelrrrrcCs.|dur tjj|g|dStjj|g||dS)Nrr)rr dim_IntList IntList_out)rwrrr'r'r( sum_defaultsrcCsBt|tjs|S|durtj|tt|Stj||gSr5) r-r?rrrdimsrSrSrO)rwrOr'r'r(squeeze_defaults rcs`tfddtt|jD}|jtjkrtjnd}|jd|d|d}||| |j|fS)Nc3s|] }|kr|VqdSr5r'rIr{r'r(rrz)_weight_norm_interface..r T)rr) rrSr`rrr?r-rrr7)rwrRrOkeep_dim norm_dtyperr'r{r(_weight_norm_interfaces r assume_uniqueinvertcCst|tjstj||jd}t|tjs"|rt||St||Sddlm}|| dt | dkr>t |||dSt ||||dS)Nrkrrg$@g(\?rr) r-r?rrrlnerrQrrr isin_default isin_sorting)elements test_elementsrrrr'r'r(isins     r)rcCsP|durtj|tj|jd}n tj||tj|jd}||k|j}|S)Nr)rrrl)r?randrrrlr7r)rwrraw_pr4r'r'r( bernoulli$srrcCs|dkrtj|tjdS|jdkr||k}|r|S|S|jd|j}||}ttd|j dd}||kj |d}|rD|S|S)Nrrr rRrr{) rr? empty_likerrr[rrtrrSrF)rrrr expanded_elem_shaper0rOr'r'r(r7s   rcCs|}|}|rMt||g}tj|dd\}}|dd|ddk} t| ddgd} |r5| } t| } | d|| } | d| |j St|\} } t | |} t | | k| d}| ||k}|rq|n|}| |j S)NT)stablerrRrF) rr?rsortr* logical_notrrrrr searchsortedrf)rrrr elements_flattest_elements_flat all_elementssorted_elements sorted_orderduplicate_maskrysorted_test_elementsrUrxtest_idxcmpr'r'r(rDs$    rcCs|d}||SrQ)r)rwr{ flattenedr'r'r(takeas rcCs2|durtj}|tjkrt|}tj||j|dSr)r?r<preserve_formatrrresizer)rwrr9r'r'r( resize_ashs  r ceil_modecCsRtrtS|jjdkrtS|d}|d} |d} |d} |dk} | s9|d}|d}|d}|d} |d}|jtj tj fv}|rQtj n|j}tj | ||| ||jd}| | || | }| | || | }|r~|tj }|d||}| | ||| }|r||j}t|}|j|d}| s|d}|S) u Decomposition of max_pool2d_with_indices_backward using scatter_add. This replaces the native implementation with a high-level decomposition that uses scatter_add for gradient accumulation. The scatter-based approach provides automatic optimization opportunities for Inductor and handles all pooling configurations without requiring specialized fallback paths. Algorithm: For each output gradient position, use the corresponding index from the forward pass to scatter the gradient to the input position. When multiple output positions select the same input position as max, scatter_add automatically accumulates their gradients. Complexity: O(B * C * H_out * W_out) Independent of kernel size, unlike traditional O(B * C * H_in * W_in * K²) approaches that iterate over input positions and kernel windows. Known Limitations: - FP16/BF16: Uses FP32 accumulation internally to preserve precision when many gradients accumulate to the same position (overlapping pooling windows). This adds slight overhead but ensures numerical stability. - Deterministic mode: Falls back to native implementation to ensure consistent results across runs Args: grad_output: Gradient w.r.t. pooling output [B, C, H_out, W_out] self: Original input tensor (for shape) [B, C, H_in, W_in] kernel_size: Pooling kernel size stride: Pooling stride padding: Pooling padding dilation: Pooling dilation ceil_mode: Whether to use ceil for output size calculation indices: Indices from forward pass (per-channel linear positions) Returns: Gradient w.r.t. input [B, C, H_in, W_in] mpsr:rRr*rrrr8)r?$are_deterministic_algorithms_enabledNotImplementedrlrSrrOrTrrr-rrrr7 scatter_addrDrrr)rirwrrTrrrrw in_heightin_width out_height out_width is_batchedrchannelsuse_fp32_accum accum_dtypegrad_input_flatgrad_output_flatrLrWr9r'r'r( max_pool2d_with_indices_backwardqsL2                 r window_lengthcCstjj|d||||dS)ahann_window(window_length, *, dtype=None, layout=None, device=None, pin_memory=False) -> Tensor Returns a Hann window of size :attr:`window_length` with ``periodic=True``. Equivalent to :func:`torch.hann_window` with ``periodic=True``. Args: window_length (int): the size of returned window. Keyword args: dtype (:class:`torch.dtype`, optional): desired dtype. Default: global default. layout (:class:`torch.layout`, optional): desired layout. Default: ``torch.strided``. device (:class:`torch.device`, optional): desired device. Default: current device. pin_memory (bool, optional): if ``True``, pins the returned tensor. Default: ``False``. Tr^)r hann_windowperiodic)rrrrlrr'r'r(rsrrc Cs|dur|nt}|dkrtjd||||dS|dkr&tjd||||dSt|}|r1|dn|}tj|||||d}|dtj|d}t|}|dd }|r\| dd|n|} | |S) a@hann_window(window_length, periodic=True, *, dtype=None, layout=None, device=None, pin_memory=False) -> Tensor Returns a Hann window of size :attr:`window_length`. .. math:: w[n] = 0.5 - 0.5 \cos\!\left(\frac{2\pi n}{N-1}\right) where :math:`N` is ``window_length + 1`` when ``periodic=True`` (for spectral analysis), or ``window_length`` when ``periodic=False`` (symmetric window). Low-precision dtypes (``bfloat16``, ``float16``) are computed in ``float32`` then cast. Args: window_length (int): the size of returned window. periodic (bool, optional): if ``True``, returns a periodic window for use with STFT. Default: ``True``. Keyword args: dtype (:class:`torch.dtype`, optional): desired dtype. Default: global default. layout (:class:`torch.layout`, optional): desired layout. Default: ``torch.strided``. device (:class:`torch.device`, optional): desired device. Default: current device. pin_memory (bool, optional): if ``True``, pins the returned tensor. Default: ``False``. Nrrr^rr rrr) r?get_default_dtyperrqrDrrppicosrr7) rrrrrlr compute_dtyperr'windowr'r'r(hann_window_periodic s."      r num_classescCsr|dkrt|d}tjt|dkdtjt||kd|dtj ||j |j dk tj S)NrRrrz+one_hot: Class values must be non-negative.z7one_hot: Class values must be smaller than num_classes.r)rhrrQr _assert_asyncmsgr?rrTrprrlr7r)rwrr'r'r(one_hotbs  r)FF)rr5r)rNNr)rRFFrrr)rrF)Fr)rrdN)rrN)Fr)NNN)rrFT)rrFr)rFr)rR(rLrSrrGrUcollections.abcrr contextlibrenumrrrrr typingr r r?torch._meta_registrations torch._primsrrtorch._prims_common _prims_commonrDtorch.nn.functionalr5rrr r r torch._decomprrerrrrrrtorch._prims_common.wrappersrrrr torch.utilsrrBtorch.utils._pytreerr DispatchKeyrrSstr__annotations___opsrrrrrrrNrcompute_only_pw_cast_for_opmathpw_cast_for_opmath"pw_cast_for_opmath_non_tensor_argsrqpw_cast_for_int_to_realrhrVr^r`rhrrsr~rryrrrrrrrrrrrrrrrrrrrrrrr%rxr _safe_softmaxrrrrdrWrrrr rr r)r+r/r1r2r3r5rGrNslicerarirMr}rrrrrrrrrrrpy_implCompositeImplicitAutogradAutogradrrrrr&rr2r4r6r9r:rArDrFr@rTrMr\rarfrirrrrrrrrrr unsafe_chunkrrrno_statsrrrrrr_fused_dropoutrrrlr9rKlift lift_freshrrrrrrrr_adaptive_avg_pool2drBrMr]rarhrirfrrrrrr Generatorrrrrrgrrr_upsample_nearest_exact1drrrrrrrrrrrrrrrrrnn_tanhrrrnn_relur rrrrrr$r2lstmr4r5r<r=grur>r?rBrGrIrJrKrLrQrRrPrNrSrVr[rarMrxryrrrrurrrrrrrrrZrrrrrrrrrhrrr=upsample_bicubic2dr0r1reflection_pad1dreflection_pad2dreflection_pad3dr:replication_pad1dreplication_pad2dreplication_pad3dr<r8reflection_pad1d_backwardreflection_pad2d_backwardreflection_pad3d_backwardrYr[r]rprrrarIrcrfrnry+_scaled_dot_product_flash_attention_for_cpurrrrrrrrrOrrrrrrrrrrr periodic_outraddbmm_addbmmaddmm_addmv_baddbmm_fill_gelu_rd hardswish_ hardtanh_hardtanh hardsigmoid___iand____and__ __ilshift__ __lshift__rsrt index_reduce_ index_reduce__ior____or__ __irshift__ __rshift____ixor____xor__ldexp_ leaky_relu_ leaky_relulogit_logitrelu_rerenorm_renormround_rsscatter_r scatter_add_rscatter_reduce_scatter_reducesilu_rr'r'r'r(s>            *                  & 1'     6  '"  P`      %    (   , 300     W       U       6K  U  #            1           m    %   $g-7  ("$$                &               .  2 )    @ 2              M"      6     ( .$$ * *   *  |  W   0  0      W     ,   <  D    "      v  !  9(