C/C++ API#
The C/C++ API allows you to access the custom kernels defined in libtransformer_engine.so library directly from C/C++, without Python.
Headers
- transformer_engine.h
NVTETensorNVTEQuantizationConfigNVTEGroupedTensorNVTEDTypeNVTETensorParamNVTEScalingModeNVTENVFP44Over6ModeNVTEQuantizationConfigAttributekNVTEQuantizationConfigForcePow2ScaleskNVTEQuantizationConfigAmaxEpsilonkNVTEQuantizationConfigNoopTensorkNVTEQuantizationConfigFloat8BlockScaleTensorFormatkNVTEQuantizationConfigRNGStatekNVTEQuantizationConfigNVFP42DQuantizationkNVTEQuantizationConfigStochasticRoundingkNVTEQuantizationConfigUseFastMathkNVTEQuantizationConfigNVFP44Over6ModekNVTEQuantizationConfigNVFP44Over6ErrUseFastMathkNVTEQuantizationConfigMXFP82DQuantizationkNVTEQuantizationConfigNumAttributes
NVTEGroupedTensorParamkNVTEGroupedRowwiseDatakNVTEGroupedColumnwiseDatakNVTEGroupedScalekNVTEGroupedAmaxkNVTEGroupedRowwiseScaleInvkNVTEGroupedColumnwiseScaleInvkNVTEGroupedColumnwiseAmaxkNVTEGroupedFirstDimskNVTEGroupedLastDimskNVTEGroupedTensorOffsetskNVTEGroupedWithGEMMSwizzledScaleskNVTENumGroupedTensorParams
nvte_create_tensor()nvte_create_tensors()nvte_destroy_tensor()nvte_destroy_tensors()nvte_tensor_data()nvte_tensor_columnwise_data()nvte_make_shape()nvte_tensor_shape()nvte_tensor_columnwise_shape()nvte_tensor_ndims()nvte_tensor_size()nvte_tensor_size_bytes()nvte_tensor_numel()nvte_tensor_element_size()nvte_tensor_element_size_bits()nvte_tensor_type()nvte_tensor_amax()nvte_tensor_scale()nvte_tensor_scale_inv()nvte_tensor_scale_inv_shape()nvte_zero_tensor()nvte_set_tensor_param()nvte_get_tensor_param()nvte_set_tensor_param_v2()nvte_get_tensor_param_v2()nvte_tensor_scaling_mode()nvte_tensor_pack_create()nvte_tensor_pack_destroy()nvte_create_quantization_config()nvte_get_quantization_config_attribute()nvte_set_quantization_config_attribute()nvte_destroy_quantization_config()nvte_is_non_tn_fp8_gemm_supported()nvte_memset()nvte_splits_to_offsets()nvte_splits_to_offsets_2d()nvte_splits_to_offsets_multi()nvte_create_grouped_tensor()nvte_destroy_grouped_tensor()nvte_set_grouped_tensor_param()nvte_get_grouped_tensor_param()nvte_grouped_tensor_num_tensors()nvte_grouped_tensor_type()nvte_grouped_tensor_scaling_mode()nvte_get_grouped_tensor_logical_shape()NVTEShapeNVTEBasicTensorNVTETensorPacktransformer_engineDTypeFloat8BlockScaleTensorFormatis_fp8_dtype()is_fp4_dtype()is_high_precision_dtype()GroupedTensorWrapperGroupedTensorWrapper()GroupedTensorWrapper()~GroupedTensorWrapper()operator=()GroupedTensorWrapper()GroupedTensorWrapper()operator=()set_parameter()set_rowwise_data()set_columnwise_data()set_scale()set_amax()set_rowwise_scale_inv()set_columnwise_scale_inv()set_columnwise_amax()set_first_dims()set_last_dims()set_tensor_offsets()set_with_gemm_swizzled_scales()get_parameter()get_rowwise_data()get_columnwise_data()get_scale()get_amax()get_rowwise_scale_inv()get_columnwise_scale_inv()get_columnwise_amax()get_first_dims()get_last_dims()get_tensor_offsets()get_with_gemm_swizzled_scales()data()num_tensors()dtype()scaling_mode()logical_shape()defaultDatadefaultShapeemptyShapeconvertShape()convertShape()tensor_
MultiTensorWrapperQuantizationConfigWrapperQuantizationConfigWrapper()QuantizationConfigWrapper()operator=()QuantizationConfigWrapper()operator=()~QuantizationConfigWrapper()operator NVTEQuantizationConfig()set_force_pow_2_scales()set_amax_epsilon()set_noop_tensor()set_float8_block_scale_tensor_format()set_rng_state()set_nvfp4_2d_quantization()set_mxfp8_2d_quantization()set_stochastic_rounding()set_use_fast_math()set_nvfp4_4over6_mode()set_nvfp4_4over6_err_use_fast_math()config_
TensorWrapperTensorWrapper()TensorWrapper()TensorWrapper()~TensorWrapper()operator=()TensorWrapper()TensorWrapper()operator=()set_parameter()set_rowwise_data()set_columnwise_data()set_scale()set_amax()set_rowwise_scale_inv()set_columnwise_scale_inv()set_columnwise_amax()set_with_gemm_swizzled_scales()set_row_scaled_nvfp4()set_nvfp4_e4m3_max()get_parameter()get_rowwise_data()get_columnwise_data()get_scale()get_amax()get_rowwise_scale_inv()get_columnwise_scale_inv()get_columnwise_amax()get_with_gemm_swizzled_scales()get_row_scaled_nvfp4()get_nvfp4_e4m3_max()data()shape()columnwise_shape()size()ndim()numel()element_size()element_size_bits()bytes()dtype()dptr()columnwise_dptr()amax()scale()scale_inv()scale_inv_shape()scaling_mode()zero_()defaultDatadefaultShapeemptyShapeconvertShape()convertShape()tensor_
- activation.h
NVTE_Activation_Typenvte_gelu()nvte_group_gelu()nvte_silu()nvte_group_silu()nvte_group_scaled_swiglu()nvte_group_scaled_clamped_swiglu()nvte_relu()nvte_group_relu()nvte_qgelu()nvte_group_qgelu()nvte_srelu()nvte_group_srelu()nvte_dgelu()nvte_group_dgelu()nvte_dsilu()nvte_group_dsilu()nvte_drelu()nvte_group_drelu()nvte_dqgelu()nvte_group_dqgelu()nvte_dsrelu()nvte_group_dsrelu()nvte_glu()nvte_dglu()nvte_geglu()nvte_swiglu()nvte_situglu()nvte_clamped_swiglu()nvte_clamped_swiglu_v2()nvte_scaled_swiglu()nvte_scaled_situglu()nvte_scaled_clamped_swiglu()nvte_reglu()nvte_qgeglu()nvte_sreglu()nvte_dgeglu()nvte_dswiglu()nvte_dsituglu()nvte_clamped_dswiglu()nvte_clamped_dswiglu_v2()nvte_scaled_dswiglu()nvte_scaled_dsituglu()nvte_scaled_clamped_dswiglu()nvte_dreglu()nvte_dqgeglu()nvte_dsreglu()nvte_scaled_srelu()nvte_scaled_dsrelu()
- cast_transpose_noop.h
- cast.h
nvte_quantize()nvte_group_quantize()nvte_quantize_noop()nvte_quantize_v2()nvte_quantize_dbias()nvte_group_quantize_dbias()nvte_quantize_dbias_dgelu()nvte_group_quantize_dbias_dgelu()nvte_quantize_dbias_dsilu()nvte_group_quantize_dbias_dsilu()nvte_quantize_dbias_drelu()nvte_group_quantize_dbias_drelu()nvte_quantize_dbias_dqgelu()nvte_group_quantize_dbias_dqgelu()nvte_quantize_dbias_dsrelu()nvte_group_quantize_dbias_dsrelu()nvte_dequantize()nvte_group_dequantize()nvte_group_requantize()nvte_multi_tensor_quantize()nvte_group_nvfp4_quantize_with_amax()
- cudnn.h
- fused_attn.h
NVTE_QKV_LayoutNVTE_SB3HDNVTE_SBH3DNVTE_SBHD_SB2HDNVTE_SBHD_SBH2DNVTE_SBHD_SBHD_SBHDNVTE_BS3HDNVTE_BSH3DNVTE_BSHD_BS2HDNVTE_BSHD_BSH2DNVTE_BSHD_BSHD_BSHDNVTE_T3HDNVTE_TH3DNVTE_THD_T2HDNVTE_THD_TH2DNVTE_THD_THD_THDNVTE_SBHD_BSHD_BSHDNVTE_BSHD_SBHD_SBHDNVTE_THD_BSHD_BSHDNVTE_THD_SBHD_SBHDNVTE_Paged_KV_BSHD_BSHD_BSHDNVTE_Paged_KV_BSHD_SBHD_SBHDNVTE_Paged_KV_SBHD_BSHD_BSHDNVTE_Paged_KV_SBHD_SBHD_SBHDNVTE_Paged_KV_THD_BSHD_BSHDNVTE_Paged_KV_THD_SBHD_SBHDNVTE_BHSD_BHSD_BHSDNVTE_QKV_Layout_NOT_SET
NVTE_QKV_Layout_GroupNVTE_QKV_FormatNVTE_Bias_TypeNVTE_Mask_TypeNVTE_Softmax_TypeNVTE_Fused_Attn_Backendnvte_get_qkv_layout_group()nvte_get_qkv_format()nvte_get_q_format()nvte_get_kv_format()nvte_get_fused_attn_backend()nvte_fused_attn_fwd()nvte_fused_attn_bwd()nvte_populate_rng_state_async()nvte_get_runtime_num_segments()nvte_extract_seed_and_offset()nvte_copy_to_kv_cache()nvte_cp_thd_read_half_tensor()nvte_cp_thd_second_half_lse_correction()nvte_cp_thd_read_second_half_lse()nvte_cp_thd_out_correction()nvte_cp_thd_grad_correction()nvte_cp_thd_get_partitioned_indices()nvte_thd_sequence_order_to_cp_rank_order()nvte_thd_cp_rank_order_to_sequence_order()nvte_thd_copy_valid_tokens_from_per_split_to_rank_local()nvte_convert_thd_to_bshd()nvte_convert_bshd_to_thd()nvte_prepare_flash_attn_fwd()nvte_prepare_flash_attn_bwd()nvte_multi_tensor_transpose_to_bhsd()nvte_multi_tensor_pad_last_dim()AttentionShape
- fused_rope.h
- gemm.h
NVTEMatmulConfigNVTEGroupedMatmulConfigNVTEMatmulConfigAttributeNVTEGroupedMatmulConfigAttributenvte_create_matmul_config()nvte_get_matmul_config_attribute()nvte_set_matmul_config_attribute()nvte_destroy_matmul_config()nvte_create_grouped_matmul_config()nvte_get_grouped_matmul_config_attribute()nvte_set_grouped_matmul_config_attribute()nvte_destroy_grouped_matmul_config()nvte_cublas_gemm()nvte_cublas_gemm_v2()nvte_cublas_gemm_scaled()nvte_cublas_atomic_gemm()nvte_multi_tensor_gemm()nvte_get_grouped_gemm_setup_workspace_size()nvte_convert_int32_to_int64()nvte_convert_int32_to_int64_with_multiplier()nvte_compute_grouped_tensor_offsets()nvte_grouped_gemm()nvte_grouped_gemm_with_discrete_inputA()nvte_grouped_gemm_with_discrete_out()nvte_grouped_bias_add()nvte_grouped_scaled_bias_add()transformer_enginenvte_cublas_handle_init()GroupedMatmulConfigWrapperMatmulConfigWrapperMatmulConfigWrapper()MatmulConfigWrapper()operator=()MatmulConfigWrapper()operator=()~MatmulConfigWrapper()operator NVTEMatmulConfig()set_bias_tensor()set_dbias_tensor()set_with_gelu_epilogue()set_with_dgelu_epilogue()set_epilogue_aux_tensor()set_use_split_accumulator()set_sm_count()config_
- multi_tensor.h
nvte_multi_tensor_l2norm_cuda()nvte_multi_tensor_unscale_l2norm_cuda()nvte_multi_tensor_adam_cuda()nvte_multi_tensor_adam_param_remainder_cuda()nvte_multi_tensor_adam_fp8_cuda()nvte_multi_tensor_adam_capturable_cuda()nvte_multi_tensor_adam_capturable_master_cuda()nvte_multi_tensor_sgd_cuda()nvte_multi_tensor_scale_cuda()nvte_multi_tensor_scale_tensor_cuda()nvte_multi_tensor_compute_scale_and_scale_inv_cuda()nvte_multi_tensor_compute_scale_inv_e8m0_cuda()nvte_group_amax()nvte_group_amax_graph_safe()
- normalization.h
- padding.h
- permutation.h
- recipe.h
nvte_delayed_scaling_recipe_amax_and_scale_update()nvte_delayed_scaling_recipe_amax_and_scale_update_after_reduction()nvte_compute_amax()nvte_compute_amax_with_config()nvte_group_compute_amax_with_config()nvte_compute_scale_from_amax()nvte_group_compute_scale_from_amax()nvte_fp8_block_scaling_compute_partial_amax()nvte_fp8_block_scaling_partial_cast()nvte_mxfp8_scaling_compute_partial_amax()nvte_mxfp8_scaling_partial_cast()nvte_nvfp4_compute_per_tensor_scale()nvte_nvfp4_2d_compute_partial_amax()nvte_nvfp4_2d_partial_cast()nvte_nvfp4_expand_scale_to_fp8()nvte_nvfp4_compute_per_block_scale()nvte_nvfp4_fused_scale()nvte_nvfp4_compute_global_scale()
- softmax.h
nvte_scaled_softmax_forward()nvte_scaled_softmax_backward()nvte_scaled_masked_softmax_forward()nvte_scaled_masked_softmax_backward()nvte_scaled_upper_triang_masked_softmax_forward()nvte_scaled_upper_triang_masked_softmax_backward()nvte_scaled_aligned_causal_masked_softmax_forward()nvte_scaled_aligned_causal_masked_softmax_backward()
- swizzle.h
nvte_swizzle_scaling_factors()nvte_multi_tensor_swizzle_scaling_factors()nvte_multi_tensor_swizzle_scaling_factors_unchecked()nvte_unswizzle_scaling_factors()nvte_multi_tensor_unswizzle_scaling_factors()nvte_swizzle_block_scaling_to_mxfp8_scaling_factors()nvte_swizzle_grouped_scaling_factors()nvte_unswizzle_grouped_scaling_factors()
- transpose.h
nvte_cast_transpose()nvte_transpose()nvte_cast_transpose_dbias()nvte_fp8_transpose_dbias()nvte_multi_cast_transpose()nvte_cast_transpose_dbias_dgelu()nvte_cast_transpose_dbias_dsilu()nvte_cast_transpose_dbias_drelu()nvte_cast_transpose_dbias_dqgelu()nvte_cast_transpose_dbias_dsrelu()nvte_dgeglu_cast_transpose()nvte_dswiglu_cast_transpose()nvte_dreglu_cast_transpose()nvte_dqgeglu_cast_transpose()nvte_dsreglu_cast_transpose()nvte_swap_first_dims()nvte_nvfp4_data_transpose()nvte_nvfp4_scale_transpose()