已合并
translation #5734
molly123321创建于 25 天前
translation #5734
已合并
共 529 个文件变更+46658-0
| @@ -0,0 +1,79 @@ | |||
| 1 | +# OpPlugin Security Statement | ||
| 2 | + | ||
| 3 | +## System Security Hardening | ||
| 4 | + | ||
| 5 | +You are advised to enable **address space layout randomization** (ASLR) (level 2) in the system. Run the following command to enable it: | ||
| 6 | + | ||
| 7 | + echo 2 > /proc/sys/kernel/randomize_va_space | ||
| 8 | + | ||
| 9 | +## Recommended Running Users | ||
| 10 | + | ||
| 11 | +The execution of OpPlugin depends on TorchNPU. For details, see [Running User Recommendations](https://gitcode.com/Ascend/pytorch/blob/v2.7.1-26.1.0/docs/en/SECURITYNOTE.md#running-user-recommendations) in the TorchNPU repository. | ||
| 12 | + | ||
| 13 | +## File Permission Control | ||
| 14 | + | ||
| 15 | +1. Control file and directory permissions during installation and use. Secure the permissions by referring to [Recommended Maximum Scenario Permissions for Files and Folders](#11111). To save installation or uninstallation logs, add `--log \<FILE>` to the end of your command, and restrict permissions on the `\<FILE>` file and directory. | ||
| 16 | + | ||
| 17 | +2. Set the running system `umask` value to `0027` or higher on both the host and the container. This setting ensures that the default maximum permission is `750` for new folders and `640` for new files. | ||
| 18 | + | ||
| 19 | +<h3 id="11111">Recommended Maximum Scenario Permissions for Files and Folders</h3> | ||
| 20 | + | ||
| 21 | +| Type | Maximum Linux Permission | | ||
| 22 | +|----------------------------------- |-----------------------| | ||
| 23 | +| Home directory | 750 (rwxr-x---) | | ||
| 24 | +| Program files (including scripts and library files) | 550 (r-xr-x---) | | ||
| 25 | +| Program file directory | 550 (r-xr-x---) | | ||
| 26 | +| Configuration files | 640 (rw-r-----) | | ||
| 27 | +| Configuration file directory | 750 (rwxr-x---) | | ||
| 28 | +| Log files (recorded or archived) | 440 (r--r-----) | | ||
| 29 | +| Log files (being recorded) | 640 (rw-r-----) | | ||
| 30 | +| Log file directory | 750 (rwxr-x---) | | ||
| 31 | +| Debug files | 640 (rw-r-----) | | ||
| 32 | +| Debug file directory | 750 (rwxr-x---) | | ||
| 33 | +| Temporary file directory | 750 (rwxr-x---) | | ||
| 34 | +| Maintenance and upgrade file directory | 770 (rwxrwx---) | | ||
| 35 | +| Service data files | 640 (rw-r-----) | | ||
| 36 | +| Service data file directory | 750 (rwxr-x---) | | ||
| 37 | +| Key components, private keys, certificates, and ciphertext file directory | 700 (rwx------) | | ||
| 38 | +| Key components, private keys, certificates, and ciphertext files | 600 (rw-------) | | ||
| 39 | +| APIs and scripts for encryption and decryption | 500 (r-x------) | | ||
| 40 | + | ||
| 41 | +## Debugging Tool Statement | ||
| 42 | + | ||
| 43 | +The execution of OpPlugin depends on TorchNPU. For details, see [Debugging Tool Statement](https://gitcode.com/Ascend/pytorch/blob/v2.7.1-26.1.0/docs/en/SECURITYNOTE.md#debugging-tool-statement) in the TorchNPU repository. | ||
| 44 | + | ||
| 45 | +## Data Security Statement | ||
| 46 | + | ||
| 47 | +The execution of OpPlugin depends on TorchNPU. For details, see [Data Security Statement](https://gitcode.com/Ascend/pytorch/blob/v2.7.1-26.1.0/docs/en/SECURITYNOTE.md#data-security-statement) in the TorchNPU repository. | ||
| 48 | + | ||
| 49 | +## Build Security Statement | ||
| 50 | + | ||
| 51 | +The execution of OpPlugin depends on TorchNPU. For details, see [Build Security Statement](https://gitcode.com/Ascend/pytorch/blob/v2.7.1-26.1.0/docs/en/SECURITYNOTE.md#build-security-statement) in the TorchNPU repository. | ||
| 52 | + | ||
| 53 | +## Runtime Security Statement | ||
| 54 | + | ||
| 55 | +The execution of OpPlugin depends on TorchNPU. For details, see [Runtime Security Statement](https://gitcode.com/Ascend/pytorch/blob/v2.7.1-26.1.0/docs/en/SECURITYNOTE.md#runtime-security-statement) in the TorchNPU repository. | ||
| 56 | + | ||
| 57 | +## Public IP Address Statement | ||
| 58 | + | ||
| 59 | +Public IP addresses are used in the configuration files and scripts of OpPlugin. For details, see [Public IP Addresses](#public-ip-addresses). | ||
| 60 | + | ||
| 61 | +### Public IP Addresses | ||
| 62 | + | ||
| 63 | +| Type | Open-Source Code Address | File | Public IP Address/Public URL/Domain Name/Email Address | Description | | ||
| 64 | +|------------------------|-------------------------|-------------------------|------------------------------------------------------------------------------------------------------------|-------------------------| | ||
| 65 | +| Development introduction | N/A | ci\build.sh | [https://gitcode.com/ascend/pytorch.git](https://gitcode.com/ascend/pytorch.git) | The build script pulls code from the TorchNPU repository address to execute the build process. | | ||
| 66 | +| Development introduction | N/A | ci\exec_ut.sh | [https://gitcode.com/ascend/pytorch.git](https://gitcode.com/ascend/pytorch.git) | The `UT` script pulls code from the TorchNPU repository address to perform `UT` tests. | | ||
| 67 | +| Open-source code introduction|pytorch\aten\src\ATen\native\TensorCompare.cpp | op_plugin\ops\opapi\IsInKernelNpuOpApi.cpp | [https://github.com/numpy/numpy/blob/fb215c76967739268de71aa4bda55dd1b062bc2e/numpy/lib/arraysetops.py#L575](https://github.com/numpy/numpy/blob/fb215c76967739268de71aa4bda55dd1b062bc2e/numpy/lib/arraysetops.py#L575) | The algorithm implementation references this NumPy source code URL.| | ||
| 68 | + | ||
| 69 | +## Public API Statement | ||
| 70 | + | ||
| 71 | +The execution of OpPlugin depends on TorchNPU, and no public API is provided. | ||
| 72 | + | ||
| 73 | +## Communication Security Hardening | ||
| 74 | + | ||
| 75 | +The execution of OpPlugin depends on TorchNPU. For details, see [Communication Security Hardening](https://gitcode.com/Ascend/pytorch/blob/v2.7.1-26.1.0/docs/en/SECURITYNOTE.md#communication-security-hardening) in the TorchNPU repository. | ||
| 76 | + | ||
| 77 | +## Communication Matrix | ||
| 78 | + | ||
| 79 | +The execution of OpPlugin depends on TorchNPU. For details, see [Communication Matrix](https://gitcode.com/Ascend/pytorch/blob/v2.7.1-26.1.0/docs/en/SECURITYNOTE.md#communication-matrix) in the TorchNPU repository. | ||
| @@ -0,0 +1 @@ | |||
| 1 | +# Python APIs | ||
| @@ -0,0 +1,3 @@ | |||
| 1 | +# Appendix | ||
| 2 | + | ||
| 3 | +- **[Example of Adding a Binary Blacklist](blacklist.md)** | ||
| @@ -0,0 +1,25 @@ | |||
| 1 | +# Example of Adding a Binary Blocklist | ||
| 2 | + | ||
| 3 | +See the following examples to set the `NPU_FUZZY_COMPILE_BLACKLIST` option to add a binary blocklist. | ||
| 4 | + | ||
| 5 | +Example for a single operator: | ||
| 6 | + | ||
| 7 | +```python | ||
| 8 | +import torch | ||
| 9 | +import torch_npu | ||
| 10 | + | ||
| 11 | +option = {} | ||
| 12 | +option['NPU_FUZZY_COMPILE_BLACKLIST'] = "DynamicGRUV2" # Configure based on the actual situation | ||
| 13 | +torch.npu.set_option(option) | ||
| 14 | +``` | ||
| 15 | + | ||
| 16 | +Example for multiple operators: | ||
| 17 | + | ||
| 18 | +```python | ||
| 19 | +import torch | ||
| 20 | +import torch_npu | ||
| 21 | + | ||
| 22 | +option = {} | ||
| 23 | +option['NPU_FUZZY_COMPILE_BLACKLIST'] = "DynamicGRUV2,DynamicRNN" # Configure based on the actual situation | ||
| 24 | +torch.npu.set_option(option) | ||
| 25 | +``` | ||
| @@ -0,0 +1,22 @@ | |||
| 1 | +# (beta) at::Device | ||
| 2 | + | ||
| 3 | +## Prototype | ||
| 4 | + | ||
| 5 | +```cpp | ||
| 6 | +at::Device(const std::string &device_string) | ||
| 7 | +``` | ||
| 8 | + | ||
| 9 | +## Function | ||
| 10 | + | ||
| 11 | +After TorchNPU is installed, the `NPU` field is supported for the Device type, allowing a device to be specified using a string description. | ||
| 12 | + | ||
| 13 | +## Parameters | ||
| 14 | + | ||
| 15 | +**`device_string`** (`string`): Input string that must be provided in the format `(npu)[:<device-index>]`, where `npu` specifies the device type and `<device-index>` (optional) specifies the device index. | ||
| 16 | + | ||
| 17 | +## Supported Products | ||
| 18 | + | ||
| 19 | +- <term>Atlas training products</term> | ||
| 20 | +- <term>Atlas A2 training products</term> | ||
| 21 | +- <term>Atlas A3 training products</term> | ||
| 22 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,26 @@ | |||
| 1 | +# (beta) at_npu::detail::createNPUGenerator | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\aten\NPUGeneratorImpl.h | ||
| 6 | + | ||
| 7 | +## Prototype | ||
| 8 | + | ||
| 9 | +```cpp | ||
| 10 | +at::Generator at_npu::detail::createNPUGenerator(c10::DeviceIndex device_index = -1) | ||
| 11 | +``` | ||
| 12 | + | ||
| 13 | +## Function | ||
| 14 | + | ||
| 15 | +Creates the default generator for an NPU device. The return value type is `Generator`, which is identical to `at::Generator at::cuda::detail::createCUDAGenerator(c10::DeviceIndex _device_index_ = -1)`. | ||
| 16 | + | ||
| 17 | +## Parameters | ||
| 18 | + | ||
| 19 | +`device_index` (`DeviceIndex`): NPU device ID for which the generator is created. | ||
| 20 | + | ||
| 21 | +## Supported Products | ||
| 22 | + | ||
| 23 | +- <term>Atlas training products</term> | ||
| 24 | +- <term>Atlas A2 training products</term> | ||
| 25 | +- <term>Atlas A3 training products</term> | ||
| 26 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,26 @@ | |||
| 1 | +# (beta) at_npu::detail::getDefaultNPUGenerator | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\aten\NPUGeneratorImpl.h | ||
| 6 | + | ||
| 7 | +## Prototype | ||
| 8 | + | ||
| 9 | +```cpp | ||
| 10 | +at::Generator& at_npu::detail::getDefaultNPUGenerator(c10::DeviceIndex device_index = -1) | ||
| 11 | +``` | ||
| 12 | + | ||
| 13 | +## Function | ||
| 14 | + | ||
| 15 | +Obtains the default generator for an NPU device. The return value type is `Generator`, which is identical to `at::Generator& at::cuda::detail::getDefaultCUDAGenerator(c10::DeviceIndex device_index = -1)`. | ||
| 16 | + | ||
| 17 | +## Parameters | ||
| 18 | + | ||
| 19 | +**`device_index`** (`DeviceIndex`): Required. NPU device ID whose generator is to be obtained. | ||
| 20 | + | ||
| 21 | +## Supported Products | ||
| 22 | + | ||
| 23 | +- <term>Atlas training products</term> | ||
| 24 | +- <term>Atlas A2 training products</term> | ||
| 25 | +- <term>Atlas A3 training products</term> | ||
| 26 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,32 @@ | |||
| 1 | +# (beta) at_npu::native::empty_with_format | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\core\npu\NPUFormat.h | ||
| 6 | + | ||
| 7 | +## Prototype | ||
| 8 | + | ||
| 9 | +```cpp | ||
| 10 | +at::Tensor at_npu::native::empty_with_format(c10::IntArrayRef sizes, const c10::TensorOptions& options, int64_t acl_format, bool keep_format = false) | ||
| 11 | +``` | ||
| 12 | + | ||
| 13 | +## Function | ||
| 14 | + | ||
| 15 | +Obtains an empty NPU tensor in a specified format. The return value type is `Tensor`, which represents the obtained empty tensor. | ||
| 16 | + | ||
| 17 | +## Parameters | ||
| 18 | + | ||
| 19 | +- **`sizes`** (`IntArrayRef`): Shape dimensions of the tensor to be obtained. | ||
| 20 | + | ||
| 21 | +- **`options`** (`TensorOptions`): Optional configuration attributes of the tensor, such as `dtype` or `device`. | ||
| 22 | + | ||
| 23 | +- **`acl_format`** (`int64_t`): Format of the tensor. | ||
| 24 | + | ||
| 25 | +- **`keep_format`** (`bool`): Optional. Specifies whether to enforce the requested format. Valid values are `True` (enforces the specified format) or `False` (allows the tensor format to be adjusted based on the actual operator execution requirements). | ||
| 26 | + | ||
| 27 | +## Supported Products | ||
| 28 | + | ||
| 29 | +- <term>Atlas training products</term> | ||
| 30 | +- <term>Atlas A2 training products</term> | ||
| 31 | +- <term>Atlas A3 training products</term> | ||
| 32 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,29 @@ | |||
| 1 | +# (beta) at_npu::native::get_npu_format | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\core\npu\NPUFormat.h | ||
| 6 | + | ||
| 7 | +## Prototype | ||
| 8 | + | ||
| 9 | +```cpp | ||
| 10 | +int64_t at_npu::native::get_npu_format(const at::Tensor& self) | ||
| 11 | +``` | ||
| 12 | + | ||
| 13 | +## Function | ||
| 14 | + | ||
| 15 | +Obtains the format information of an NPU tensor and returns an `int64_t` value. | ||
| 16 | + | ||
| 17 | +> [!NOTICE] | ||
| 18 | +> This API is typically used together with `empty_with_format` when allocating memory in an NPU proprietary format. | ||
| 19 | + | ||
| 20 | +## Parameters | ||
| 21 | + | ||
| 22 | +`self` (`Tensor`): Tensor whose format information is to be obtained. | ||
| 23 | + | ||
| 24 | +## Supported Products | ||
| 25 | + | ||
| 26 | +- <term>Atlas training products</term> | ||
| 27 | +- <term>Atlas A2 training products</term> | ||
| 28 | +- <term>Atlas A3 training products</term> | ||
| 29 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,26 @@ | |||
| 1 | +# (beta) at_npu::native::get_npu_storage_sizes | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\core\npu\NPUFormat.h | ||
| 6 | + | ||
| 7 | +## Prototype | ||
| 8 | + | ||
| 9 | +```cpp | ||
| 10 | +std::vector<int64_t> at_npu::native::get_npu_storage_sizes(const at::Tensor& self) | ||
| 11 | +``` | ||
| 12 | + | ||
| 13 | +## Function | ||
| 14 | + | ||
| 15 | +Obtains the memory size of an NPU tensor. The return value type is `vector<int64_t>`, which represents the allocated memory size of the NPU tensor. | ||
| 16 | + | ||
| 17 | +## Parameters | ||
| 18 | + | ||
| 19 | +**`self`** (`Tensor`): Tensor whose allocated memory size is to be obtained. | ||
| 20 | + | ||
| 21 | +## Supported Products | ||
| 22 | + | ||
| 23 | +- <term>Atlas training products</term> | ||
| 24 | +- <term>Atlas A2 training products</term> | ||
| 25 | +- <term>Atlas A3 training products</term> | ||
| 26 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,32 @@ | |||
| 1 | +# (beta) at_npu::native::npu_dropout_gen_mask | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +third_party\op-plugin\op_plugin\include\ops.h | ||
| 6 | + | ||
| 7 | +## Prototype | ||
| 8 | + | ||
| 9 | +```cpp | ||
| 10 | +at::Tensor npu_dropout_gen_mask(const at::Tensor &self, at::IntArrayRef size, double p, int64_t seed, int64_t offset, c10::optional<bool> parallel, c10::optional<bool> sync) | ||
| 11 | +``` | ||
| 12 | + | ||
| 13 | +## Function | ||
| 14 | + | ||
| 15 | +Generates a random mask based on the probability `p` during training, which is used to set elements to zero. | ||
| 16 | + | ||
| 17 | +## Parameters | ||
| 18 | + | ||
| 19 | +- **`self`** (`Tensor`): Input tensor. | ||
| 20 | +- **`size`** (`IntArrayRef`): Shape dimensions of the generated mask. | ||
| 21 | +- **`p`** (`double`): Probability of setting an element to 0. | ||
| 22 | +- **`seed`** (`int64_t`): Seed that determines the generated random number sequence. | ||
| 23 | +- **`offset`** (`int64_t`): Random number offset that controls the execution alignment position of the generated random number sequence. | ||
| 24 | +- **`parallel`** (`bool`): Optional. Specifies whether to enable parallel computation. | ||
| 25 | +- **`sync`** (`bool`): Optional. Specifies whether to enable synchronized execution. | ||
| 26 | + | ||
| 27 | +## Supported Products | ||
| 28 | + | ||
| 29 | +- <term>Atlas training products</term> | ||
| 30 | +- <term>Atlas A2 training products</term> | ||
| 31 | +- <term>Atlas A3 training products</term> | ||
| 32 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,37 @@ | |||
| 1 | +# (beta) at_npu::native::npu_format_cast | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\core\npu\NPUFormat.h | ||
| 6 | + | ||
| 7 | +## Prototype | ||
| 8 | + | ||
| 9 | +```cpp | ||
| 10 | +at::Tensor at_npu::native::npu_format_cast(const at::Tensor& self, int64_t acl_format) | ||
| 11 | +``` | ||
| 12 | + | ||
| 13 | +## Function | ||
| 14 | + | ||
| 15 | +Converts the format of an NPU tensor. The return value type is `Tensor`, which represents the converted tensor. | ||
| 16 | + | ||
| 17 | +## Parameters | ||
| 18 | + | ||
| 19 | +**`self`** (`Tensor`): Tensor whose format is to be converted. | ||
| 20 | + | ||
| 21 | +**`acl_format`** (`int64_t`): Destination format for conversion. | ||
| 22 | + | ||
| 23 | +## Constraints | ||
| 24 | + | ||
| 25 | +In <term>Ascend 950DT</term> scenarios, the following special cases are currently not supported when converting a tensor to the `FRACTAL_NZ` format: | ||
| 26 | + | ||
| 27 | +- If the `dtype` of `self` is `float16` or `bfloat16`, and the dimensions of `self` are represented as `[k, n]`, the case where `k` is 1 is not supported. | ||
| 28 | +- After calling this API to convert a tensor to the `FRACTAL_NZ` format, operations that modify the tensor, including `contiguous`, `pad`, `view`, and `slice`, are not supported. | ||
| 29 | +- If either of the last two dimensions of the shape of `self` is 1, `transpose` is not supported after converting to the `FRACTAL_NZ` format. | ||
| 30 | + | ||
| 31 | +## Supported Products | ||
| 32 | + | ||
| 33 | +- <term>Ascend 950DT</term> | ||
| 34 | +- <term>Atlas A3 training products</term> | ||
| 35 | +- <term>Atlas A2 training products</term> | ||
| 36 | +- <term>Atlas training products</term> | ||
| 37 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,22 @@ | |||
| 1 | +# (beta) c10::npu::current_device | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\libs\init_npu.h | ||
| 6 | + | ||
| 7 | +## Prototype | ||
| 8 | + | ||
| 9 | +```cpp | ||
| 10 | +c10::DeviceIndex c10::npu::current_device() | ||
| 11 | +``` | ||
| 12 | + | ||
| 13 | +## Function | ||
| 14 | + | ||
| 15 | +Obtains the NPU device ID. The return value type is `DeviceIndex`. This function is identical to `c10::DeviceIndex c10::cuda::current_device()`. | ||
| 16 | + | ||
| 17 | +## Supported Products | ||
| 18 | + | ||
| 19 | +- <term>Atlas training products</term> | ||
| 20 | +- <term>Atlas A2 training products</term> | ||
| 21 | +- <term>Atlas A3 training products</term> | ||
| 22 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,26 @@ | |||
| 1 | +# (beta) c10_npu::GetDevice | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\core\npu\NPUFunctions.h | ||
| 6 | + | ||
| 7 | +## Prototype | ||
| 8 | + | ||
| 9 | +```cpp | ||
| 10 | +aclError c10_npu::GetDevice(c10::DeviceIndex* device) | ||
| 11 | +``` | ||
| 12 | + | ||
| 13 | +## Function | ||
| 14 | + | ||
| 15 | +Obtains the NPU device ID. The return value type is `aclError`, which is identical to `cudaError_t c10::cuda::GetDevice(int* device)` in PyTorch 1.11.0. | ||
| 16 | + | ||
| 17 | +## Parameters | ||
| 18 | + | ||
| 19 | +**`device`** (`DeviceIndex`): Stores the obtained device ID. | ||
| 20 | + | ||
| 21 | +## Supported Products | ||
| 22 | + | ||
| 23 | +- <term>Atlas training products</term> | ||
| 24 | +- <term>Atlas A2 training products</term> | ||
| 25 | +- <term>Atlas A3 training products</term> | ||
| 26 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,26 @@ | |||
| 1 | +# (beta) c10_npu::SetDevice | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\core\npu\NPUFunctions.h | ||
| 6 | + | ||
| 7 | +## Prototype | ||
| 8 | + | ||
| 9 | +```cpp | ||
| 10 | +aclError c10_npu::SetDevice(c10::DeviceIndex device) | ||
| 11 | +``` | ||
| 12 | + | ||
| 13 | +## Function | ||
| 14 | + | ||
| 15 | +Sets the NPU device. It specifies the NPU device to be used by the current thread. The return value is of type `aclError`. | ||
| 16 | + | ||
| 17 | +## Parameters | ||
| 18 | + | ||
| 19 | +**`device`** (`DeviceIndex`): NPU device ID to be set. | ||
| 20 | + | ||
| 21 | +## Supported Products | ||
| 22 | + | ||
| 23 | +- <term>Atlas training products</term> | ||
| 24 | +- <term>Atlas A2 training products</term> | ||
| 25 | +- <term>Atlas A3 training products</term> | ||
| 26 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,22 @@ | |||
| 1 | +# (beta) c10_npu::c10_npu_get_error_message | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\core\npu\NPUException.h | ||
| 6 | + | ||
| 7 | +## Prototype | ||
| 8 | + | ||
| 9 | +```cpp | ||
| 10 | +const char* c10_npu::c10_npu_get_error_message() | ||
| 11 | +``` | ||
| 12 | + | ||
| 13 | +## Function | ||
| 14 | + | ||
| 15 | +Obtains error messages. The return value type is `const char *`, which represents the obtained error message string. | ||
| 16 | + | ||
| 17 | +## Supported Products | ||
| 18 | + | ||
| 19 | +- <term>Atlas training products</term> | ||
| 20 | +- <term>Atlas A2 training products</term> | ||
| 21 | +- <term>Atlas A3 training products</term> | ||
| 22 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,22 @@ | |||
| 1 | +# (beta) c10_npu::current_device | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\core\npu\NPUFunctions.h | ||
| 6 | + | ||
| 7 | +## Prototype | ||
| 8 | + | ||
| 9 | +```cpp | ||
| 10 | +c10::DeviceIndex c10_npu::current_device() | ||
| 11 | +``` | ||
| 12 | + | ||
| 13 | +## Function | ||
| 14 | + | ||
| 15 | +Obtains the NPU device ID. The return value type is `DeviceIndex`, which represents the obtained device ID. This function is identical to `c10::DeviceIndex c10::cuda::current_device()` in PyTorch 1.11.0. The main difference from `c10_npu::GetDevice` is that this function includes additional error checking. | ||
| 16 | + | ||
| 17 | +## Supported Products | ||
| 18 | + | ||
| 19 | +- <term>Atlas training products</term> | ||
| 20 | +- <term>Atlas A2 training products</term> | ||
| 21 | +- <term>Atlas A3 training products</term> | ||
| 22 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,22 @@ | |||
| 1 | +# (beta) c10_npu::device_count | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\core\npu\NPUFunctions.h | ||
| 6 | + | ||
| 7 | +## Prototype | ||
| 8 | + | ||
| 9 | +```cpp | ||
| 10 | +c10::DeviceIndex c10_npu::device_count() | ||
| 11 | +``` | ||
| 12 | + | ||
| 13 | +## Function | ||
| 14 | + | ||
| 15 | +Obtains the number of available NPUs. The return value type is `DeviceIndex`. | ||
| 16 | + | ||
| 17 | +## Supported Products | ||
| 18 | + | ||
| 19 | +- <term>Atlas training products</term> | ||
| 20 | +- <term>Atlas A2 training products</term> | ||
| 21 | +- <term>Atlas A3 training products</term> | ||
| 22 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,26 @@ | |||
| 1 | +# (beta) c10_npu::getCurrentNPUStream | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\core\npu\NPUStream.h | ||
| 6 | + | ||
| 7 | +## Prototype | ||
| 8 | + | ||
| 9 | +```cpp | ||
| 10 | +c10_npu::NPUStream c10_npu::getCurrentNPUStream(c10::DeviceIndex device_index = -1) | ||
| 11 | +``` | ||
| 12 | + | ||
| 13 | +## Function | ||
| 14 | + | ||
| 15 | +Obtains the current NPU stream through a device ID. The return value type is `NPUStream`, which is identical to `c10::cuda::CUDAStream c10::cuda::getCurrentCUDAStream(c10::DeviceIndex device_index = -1)`. | ||
| 16 | + | ||
| 17 | +## Parameters | ||
| 18 | + | ||
| 19 | +**`device_index`** (`DeviceIndex`): NPU device ID whose stream is to be obtained. | ||
| 20 | + | ||
| 21 | +## Supported Products | ||
| 22 | + | ||
| 23 | +- <term>Atlas training products</term> | ||
| 24 | +- <term>Atlas A2 training products</term> | ||
| 25 | +- <term>Atlas A3 training products</term> | ||
| 26 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,26 @@ | |||
| 1 | +# (beta) c10_npu::getDefaultNPUStream | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\core\npu\NPUStream.h | ||
| 6 | + | ||
| 7 | +## Prototype | ||
| 8 | + | ||
| 9 | +```cpp | ||
| 10 | +c10_npu::NPUStream c10_npu::getDefaultNPUStream(c10::DeviceIndex device_index = -1) | ||
| 11 | +``` | ||
| 12 | + | ||
| 13 | +## Function | ||
| 14 | + | ||
| 15 | +Obtains the default NPU stream. The return value type is `NPUStream`. Its function and usage are identical to `c10::cuda::CUDAStream c10::cuda::getDefaultCUDAStream(c10::DeviceIndex device_index = -1)`. | ||
| 16 | + | ||
| 17 | +## Parameters | ||
| 18 | + | ||
| 19 | +**`device_index`** (`DeviceIndex`): NPU device ID whose stream is to be obtained. The default value is `-1`, which specifies to use the current NPU device. | ||
| 20 | + | ||
| 21 | +## Supported Products | ||
| 22 | + | ||
| 23 | +- <term>Atlas training products</term> | ||
| 24 | +- <term>Atlas A2 training products</term> | ||
| 25 | +- <term>Atlas A3 training products</term> | ||
| 26 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,26 @@ | |||
| 1 | +# (beta) c10_npu::getNPUStreamFromPool | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\core\npu\NPUStream.h | ||
| 6 | + | ||
| 7 | +## Prototype | ||
| 8 | + | ||
| 9 | +```cpp | ||
| 10 | +c10_npu::NPUStream c10_npu::getNPUStreamFromPool(c10::DeviceIndex device = -1) | ||
| 11 | +``` | ||
| 12 | + | ||
| 13 | +## Function | ||
| 14 | + | ||
| 15 | +Obtains a new stream from the NPU stream pool. Streams are preallocated in the pool and retrieved in a round-robin manner. The return value type is `NPUStream`. | ||
| 16 | + | ||
| 17 | +## Parameters | ||
| 18 | + | ||
| 19 | +**`device`** (`DeviceIndex`): NPU device ID whose stream is to be obtained. | ||
| 20 | + | ||
| 21 | +## Supported Products | ||
| 22 | + | ||
| 23 | +- <term>Atlas training products</term> | ||
| 24 | +- <term>Atlas A2 training products</term> | ||
| 25 | +- <term>Atlas A3 training products</term> | ||
| 26 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,26 @@ | |||
| 1 | +# (beta) c10_npu::setCurrentNPUStream | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\core\npu\NPUStream.h | ||
| 6 | + | ||
| 7 | +## Prototype | ||
| 8 | + | ||
| 9 | +```cpp | ||
| 10 | +void c10_npu::setCurrentNPUStream(c10_npu::NPUStream stream) | ||
| 11 | +``` | ||
| 12 | + | ||
| 13 | +## Function | ||
| 14 | + | ||
| 15 | +Sets the current NPU stream. This function is identical to `void c10::cuda::setCurrentCUDAStream(c10::cuda::CUDAStream stream)`. | ||
| 16 | + | ||
| 17 | +## Parameters | ||
| 18 | + | ||
| 19 | +**`stream`** (`NPUStream`): NPU stream to be set. | ||
| 20 | + | ||
| 21 | +## Supported Products | ||
| 22 | + | ||
| 23 | +- <term>Atlas training products</term> | ||
| 24 | +- <term>Atlas A2 training products</term> | ||
| 25 | +- <term>Atlas A3 training products</term> | ||
| 26 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,26 @@ | |||
| 1 | +# (beta) c10_npu::set_device | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\core\npu\NPUFunctions.h | ||
| 6 | + | ||
| 7 | +## Prototype | ||
| 8 | + | ||
| 9 | +```cpp | ||
| 10 | +void c10_npu::set_device(c10::DeviceIndex device) | ||
| 11 | +``` | ||
| 12 | + | ||
| 13 | +## Function | ||
| 14 | + | ||
| 15 | +Sets the NPU device. This function is identical to `void c10::cuda::set_device(c10::DeviceIndex device)` in PyTorch 1.11.0. The main difference from `c10_npu::SetDevice` is that this function includes additional error checking. | ||
| 16 | + | ||
| 17 | +## Parameters | ||
| 18 | + | ||
| 19 | +**`device`** (`DeviceIndex`): NPU device ID to be set. | ||
| 20 | + | ||
| 21 | +## Supported Products | ||
| 22 | + | ||
| 23 | +- <term>Atlas training products</term> | ||
| 24 | +- <term>Atlas A2 training products</term> | ||
| 25 | +- <term>Atlas A3 training products</term> | ||
| 26 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,22 @@ | |||
| 1 | +# (beta) c10_npu::warn_or_error_on_sync | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\core\npu\NPUFunctions.h | ||
| 6 | + | ||
| 7 | +## Prototype | ||
| 8 | + | ||
| 9 | +```cpp | ||
| 10 | +void c10_npu::warn_or_error_on_sync() | ||
| 11 | +``` | ||
| 12 | + | ||
| 13 | +## Function | ||
| 14 | + | ||
| 15 | +Raises a warning or error during NPU synchronization, with no return value. This function raises an error or logs a warning based on the current warning level, which is identical to `void c10::cuda::warn_or_error_on_sync()`. | ||
| 16 | + | ||
| 17 | +## Supported Products | ||
| 18 | + | ||
| 19 | +- <term>Atlas training products</term> | ||
| 20 | +- <term>Atlas A2 training products</term> | ||
| 21 | +- <term>Atlas A3 training products</term> | ||
| 22 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,22 @@ | |||
| 1 | +# (beta) c10_npu::warning_state | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\core\npu\NPUFunctions.h | ||
| 6 | + | ||
| 7 | +## Prototype | ||
| 8 | + | ||
| 9 | +```cpp | ||
| 10 | +c10_npu::WarningState& c10_npu::warning_state() | ||
| 11 | +``` | ||
| 12 | + | ||
| 13 | +## Function | ||
| 14 | + | ||
| 15 | +Obtains the current runtime warning level. The return type is the `WarningState` enum class, which can be `L_DISABLED` (no warning), `L_WARN` (warning), or `L_ERROR` (error). This function is identical to `WarningState& c10::cuda::warning_state()` in PyTorch 1.11.0. | ||
| 16 | + | ||
| 17 | +## Supported Products | ||
| 18 | + | ||
| 19 | +- <term>Atlas training products</term> | ||
| 20 | +- <term>Atlas A2 training products</term> | ||
| 21 | +- <term>Atlas A3 training products</term> | ||
| 22 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,117 @@ | |||
| 1 | +# (beta) class at_npu::NPUGeneratorImpl | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\aten\NPUGeneratorImpl.h | ||
| 6 | + | ||
| 7 | +## Function | ||
| 8 | + | ||
| 9 | +Acts as a random number generator class that implements NPU device random number functionality and supports various random-dependent methods. | ||
| 10 | + | ||
| 11 | +## Member Functions | ||
| 12 | + | ||
| 13 | +- **at_npu::NPUGeneratorImpl::NPUGeneratorImpl(c10::DeviceIndex device_index = -1)** | ||
| 14 | + | ||
| 15 | + `NPUGeneratorImpl` constructor, which constructs a generator by specifying an NPU device ID. This function is identical to `at::CUDAGeneratorImpl::CUDAGeneratorImpl(c10::DeviceIndex device_index)`. | ||
| 16 | + | ||
| 17 | + **`device_index`** (`DeviceIndex`): NPU device ID. | ||
| 18 | + | ||
| 19 | +- **std::shared_ptr\<NPUGeneratorImpl> at_npu::NPUGeneratorImpl::clone()** | ||
| 20 | + | ||
| 21 | + `NPUGeneratorImpl` copy function. The return type is `std::shared_ptr<NPUGeneratorImpl>`, which represents the obtained `NPUGeneratorImpl` copy. This function is identical to `std::shared_ptr<CUDAGeneratorImpl> at::CUDAGeneratorImpl::clone()`. | ||
| 22 | + | ||
| 23 | +- **void at_npu::NPUGeneratorImpl::set_current_seed(uint64_t seed)** | ||
| 24 | + | ||
| 25 | + Sets the current random number seed for `NPUGeneratorImpl`. This function is identical to `at::CUDAGeneratorImpl::set_current_seed(uint64_t seed)`. | ||
| 26 | + | ||
| 27 | + **`seed`** (`uint64_t`): Random number seed to be set. | ||
| 28 | + | ||
| 29 | +- **void at_npu::NPUGeneratorImpl::set_offset(uint64_t offset)** | ||
| 30 | + | ||
| 31 | + Sets the offset value for `NPUGeneratorImpl`. This function is identical to `at::CUDAGeneratorImpl::set_offset(uint64_t offset)`. | ||
| 32 | + | ||
| 33 | + **`offset`** (`uint64_t`): Offset value to be set. | ||
| 34 | + | ||
| 35 | +- **uint64_t at_npu::NPUGeneratorImpl::current_seed()** | ||
| 36 | + | ||
| 37 | + Obtains the current random number seed for `NPUGeneratorImpl`. The return type is `uint64_t`, which represents the current random number seed. This function is identical to `at::CUDAGeneratorImpl::current_seed()`. | ||
| 38 | + | ||
| 39 | +- **uint64_t at_npu::NPUGeneratorImpl::get_offset()** | ||
| 40 | + | ||
| 41 | + Obtains the current offset value for `NPUGeneratorImpl`. The return type is `uint64_t`, which represents this offset value. This function is identical to `at::CUDAGeneratorImpl::get_offset()`. | ||
| 42 | + | ||
| 43 | +- **uint64_t at_npu::NPUGeneratorImpl::seed()** | ||
| 44 | + | ||
| 45 | + Updates the random number seed for `NPUGeneratorImpl`. The return type is `uint64_t`, which generates and returns a new random number seed. This function is identical to `at::CUDAGeneratorImpl::seed()`. | ||
| 46 | + | ||
| 47 | +- **void at_npu::NPUGeneratorImpl::set_state(const c10::TensorImpl& new_state)** | ||
| 48 | + | ||
| 49 | + Sets the specified state for `NPUGeneratorImpl`. This function is identical to `void at::CUDAGeneratorImpl::set_state(const c10::TensorImpl& new_state)`. | ||
| 50 | + | ||
| 51 | + **`new_state`** (`TensorImpl`): State to be set, which must be validated using `at::detail::check_rng_state`. | ||
| 52 | + | ||
| 53 | +- **c10::intrusive_ptr\<c10::TensorImpl> at_npu::NPUGeneratorImpl::get_state()** | ||
| 54 | + | ||
| 55 | + Obtains the generator state for `NPUGeneratorImpl`. The return type is `c10::intrusive_ptr<c10::TensorImpl>`, which represents the obtained generator state. This function is identical to `c10::intrusive_ptr<c10::TensorImpl> at::CUDAGeneratorImpl::get_state()`. | ||
| 56 | + | ||
| 57 | +- **void at_npu::NPUGeneratorImpl::set_philox_offset_per_thread(uint64_t offset)** | ||
| 58 | + | ||
| 59 | + Sets the Philox offset value for each thread for `NPUGeneratorImpl`, which is used for `curandStatePhilox4_32_10`. This function is identical to `void at::CUDAGeneratorImpl::set_philox_offset_per_thread(uint64_t offset)`. | ||
| 60 | + | ||
| 61 | + **`offset`** (`uint64_t`): Required. Philox offset value to be set. | ||
| 62 | + | ||
| 63 | +- **uint64_t at_npu::NPUGeneratorImpl::philox_offset_per_thread()** | ||
| 64 | + | ||
| 65 | + Obtains the Philox offset for each thread for `NPUGeneratorImpl`. The return type is `uint64_t`. This function is identical to `uint64_t at::CUDAGeneratorImpl::philox_offset_per_thread()`. | ||
| 66 | + | ||
| 67 | +- **at_npu::PhiloxNpuState at_npu::NPUGeneratorImpl::philox_npu_state(uint64_t increment)** | ||
| 68 | + | ||
| 69 | + Captures the Philox NPU state for `NPUGeneratorImpl`. The return type is `PhiloxNpuState`. This function is identical to `at::PhiloxCudaState at::CUDAGeneratorImpl::philox_cuda_state(uint64_t increment)`. | ||
| 70 | + | ||
| 71 | + **`increment`** (`uint64_t`): Required. Philox offset increment. | ||
| 72 | + | ||
| 73 | +- **std::pair<uint64_t, uint64_t> at_npu::NPUGeneratorImpl::philox_engine_inputs(uint64_t increment)** | ||
| 74 | + | ||
| 75 | + Obtains the Philox engine inputs for `NPUGeneratorImpl`. The return type is `std::pair<uint64_t, uint64_t>`, which contains the random number seed and the Philox offset value. This function is identical to `std::pair<uint64_t, uint64_t> at::CUDAGeneratorImpl::philox_engine_inputs(uint64_t increment)`. | ||
| 76 | + | ||
| 77 | + **`increment`** (`uint64_t`): Required. Philox offset increment. | ||
| 78 | + | ||
| 79 | +- **c10::DeviceType at_npu::NPUGeneratorImpl::device_type()** | ||
| 80 | + | ||
| 81 | + Obtains the device type for `NPUGeneratorImpl`. The return type is `DeviceType`. This function is identical to `c10::DeviceType at::CUDAGeneratorImpl::device_type()`. | ||
| 82 | + | ||
| 83 | +In PyTorch 2.5.1 and later, the following member functions are removed. In versions earlier than PyTorch 2.5.1, these member functions are still available: | ||
| 84 | + | ||
| 85 | +- **void at_npu::NPUGeneratorImpl::capture_prologue()** | ||
| 86 | + | ||
| 87 | + Sets `offset_extragraph` for `NPUGeneratorImpl`, which is used by `NPUGraph` to reserve a graph capture region and enable graph capture. This function is identical to `void at::CUDAGeneratorImpl::capture_prologue()`. | ||
| 88 | + | ||
| 89 | +- **uint64_t at_npu::NPUGeneratorImpl::capture_epilogue()** | ||
| 90 | + | ||
| 91 | + Ends graph capture for `NPUGeneratorImpl`. The return type is `uint64_t`, which disables graph capture and returns the reserved region size `offset_extragraph`. This function is identical to `uint64_t at::CUDAGeneratorImpl::capture_epilogue()`. | ||
| 92 | + | ||
| 93 | +In PyTorch 2.5.1 and later, the following member functions are added: | ||
| 94 | + | ||
| 95 | +- **void graphsafe_set_state(const c10::intrusive_ptr\<c10::GeneratorImpl>& state)** | ||
| 96 | + | ||
| 97 | + Sets the expected random number generator state for `aclgraph` during graph capture for `NPUGeneratorImpl`. This function is identical to `void at::CUDAGeneratorImpl::graphsafe_set_state(const c10::intrusive_ptr<c10::GeneratorImpl>& state)`. | ||
| 98 | + | ||
| 99 | + **`state`** (`c10::intrusive_ptr<c10::GeneratorImpl>`): Random number generator state. | ||
| 100 | +- **c10::intrusive_ptr\<c10::GeneratorImpl> graphsafe_get_state()** | ||
| 101 | + | ||
| 102 | + Queries the random number generator object for `aclgraph` during graph capture for `NPUGeneratorImpl`. The return value is a `c10::GeneratorImpl` object. This function is identical to `c10::intrusive_ptr<c10::GeneratorImpl> at::CUDAGeneratorImpl::graphsafe_get_state()`. | ||
| 103 | + | ||
| 104 | + The return value is a `c10::GeneratorImpl` object. | ||
| 105 | +- **void register_graph(c10_npu::NPUGraph\* graph)** | ||
| 106 | + | ||
| 107 | + Registers an `aclgraph` object with `NPUGeneratorImpl` for unified management. This function is identical to `void at::CUDAGeneratorImpl::register_graph(CUDAGraph* graph)`. | ||
| 108 | +- **void unregister_graph(c10_npu::NPUGraph\* graph)** | ||
| 109 | + | ||
| 110 | + Removes an `aclgraph` object from `NPUGeneratorImpl`, which is called when the graph object is destroyed. This function is identical to `void at::CUDAGeneratorImpl::unregister_graph(CUDAGraph* graph)`. | ||
| 111 | + | ||
| 112 | +## Supported Products | ||
| 113 | + | ||
| 114 | +- <term>Atlas training products</term> | ||
| 115 | +- <term>Atlas A2 training products</term> | ||
| 116 | +- <term>Atlas A3 training products</term> | ||
| 117 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,190 @@ | |||
| 1 | +# (beta) class at_npu::native::OpCommand | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\framework\OpCommand.h | ||
| 6 | + | ||
| 7 | +## Function | ||
| 8 | + | ||
| 9 | +Acts as a class encapsulating low-level operator calls and implements underlying operator execution on NPU devices. | ||
| 10 | + | ||
| 11 | +## Member Functions | ||
| 12 | + | ||
| 13 | +- **at_npu::native::OpCommand::OpCommand()** | ||
| 14 | + | ||
| 15 | + `OpCommand` constructor, which creates an `OpCommand` instance. | ||
| 16 | + | ||
| 17 | +- **at_npu::native::OpCommand::\~OpCommand()** | ||
| 18 | + | ||
| 19 | + `OpCommand` destructor. | ||
| 20 | + | ||
| 21 | +- **at_npu::native::OpCommand& at_npu::native::OpCommand::Name(const string& name)** | ||
| 22 | + | ||
| 23 | + Name of the operator to be executed. The return value type is `OpCommand&`. | ||
| 24 | + | ||
| 25 | + **`name`** (`string`): Required. Name of the operator to be executed. | ||
| 26 | + | ||
| 27 | +- **at_npu::native::OpCommand& at_npu::native::OpCommand::SetCustomHandler(PROC_FUNC func)** | ||
| 28 | + | ||
| 29 | + Sets a custom handling method for `OpCommand`. The return value type is `OpCommand`. | ||
| 30 | + | ||
| 31 | + **`func`** (`PROC_FUNC`): Custom handling method to be set. | ||
| 32 | + | ||
| 33 | +- **at_npu::native::OpCommand& at_npu::native::OpCommand::DynamicInputReg(DynamicInputRegFunc func, DyNumAndIndex num_and_index)** | ||
| 34 | + | ||
| 35 | + Registers a dynamic input method for `OpCommand`. The return value type is `OpCommand`. | ||
| 36 | + | ||
| 37 | + **`func`** (`DynamicInputRegFunc`): Dynamic input method to be registered. | ||
| 38 | + | ||
| 39 | + **`num_and_index`** (`DyNumAndIndex`): Identifier to be registered. | ||
| 40 | + | ||
| 41 | +- **at_npu::native::OpCommand& at_npu::native::OpCommand::Expect(UnifiedResult unified_result)** | ||
| 42 | + | ||
| 43 | + Sets the expected result format for `OpCommand`. The return value type is `OpCommand`. | ||
| 44 | + | ||
| 45 | + **`unified_result`** (`UnifiedResult`): Includes the type, shape, or definition state of the expected result. | ||
| 46 | + | ||
| 47 | +- **at_npu::native::OpCommand& at_npu::native::OpCommand::Input()** | ||
| 48 | + | ||
| 49 | + An empty input for `OpCommand`. The return value type is `OpCommand`. | ||
| 50 | + | ||
| 51 | +- **at_npu::native::OpCommand& at_npu::native::OpCommand::Input(const at::Tensor& input, const string& descName = "", const c10::optional& sensitive_format = c10::nullopt, const string& realData = "")** | ||
| 52 | + | ||
| 53 | + Tensor input for `OpCommand`, which must be contiguous. The return value type is `OpCommand`. | ||
| 54 | + | ||
| 55 | + **`input`** (`Tensor`): Input tensor, which must be contiguous. | ||
| 56 | + | ||
| 57 | + **`descName`** (`string`): Name description. | ||
| 58 | + | ||
| 59 | + **`sensitive_format`** (`aclFormat`): Specific format requirements. | ||
| 60 | + | ||
| 61 | + **`realData`** (`string`): Actual data type. | ||
| 62 | + | ||
| 63 | +- **at_npu::native::OpCommand& at_npu::native::OpCommand::InputWithoutContiguous(const at::Tensor& input, const string& descName = "", const string& realData = "")** | ||
| 64 | + | ||
| 65 | + Tensor input for `OpCommand`. Non-contiguous tensors are supported. The return value type is `OpCommand`. | ||
| 66 | + | ||
| 67 | + **`input`** (`Tensor`): Input tensor. | ||
| 68 | + | ||
| 69 | + **`descName`** (`string`): Name description. | ||
| 70 | + | ||
| 71 | + **`realData`** (`string`): Actual data type. | ||
| 72 | + | ||
| 73 | +- **template at_npu::native::OpCommand& at_npu::native::OpCommand::Input(const c10::ArrayRef& dimListRef, at::IntArrayRef realShape, at::ScalarType toType, CompileType compileType = CompileType::MEMORY_HOST_COMPILE_DEPENDENT, const string& realDtype = "", const string& descName = "")** | ||
| 74 | + | ||
| 75 | + Array input for `OpCommand`, typically provided on the CPU side, which is copied through H2D during execution. The return value type is `OpCommand`. | ||
| 76 | + | ||
| 77 | + **`dimListRef`** (`ArrayRef`): Input array. | ||
| 78 | + | ||
| 79 | + **`realShape`** (`IntArrayRef`): Input shape dimensions. | ||
| 80 | + | ||
| 81 | + **`toType`** (`ScalarType`): Target data type. | ||
| 82 | + | ||
| 83 | + **`compileType`** (`CompileType`): Compilation type. | ||
| 84 | + | ||
| 85 | + **`realDtype`** (`string`): Actual data type. | ||
| 86 | + | ||
| 87 | + **`descName`** (`string`): Name description. | ||
| 88 | + | ||
| 89 | +- **at_npu::native::OpCommand& at_npu::native::OpCommand::Input(const c10::IntArrayRef& dimListRef, at::ScalarType toType = at::kLong, CompileType compileType = CompileType::MEMORY_HOST_COMPILE_DEPENDENT, const string& realDtype = "", const string& descName = "")** | ||
| 90 | + | ||
| 91 | + Integer array input for `OpCommand`, typically provided on the CPU side, which is copied through H2D during execution. The return value type is `OpCommand`. | ||
| 92 | + | ||
| 93 | + **`dimListRef`** (`IntArrayRef`): Input integer array. | ||
| 94 | + | ||
| 95 | + **`toType`** (`ScalarType`): Target data type. | ||
| 96 | + | ||
| 97 | + **`compileType`** (`CompileType`): Compilation type. | ||
| 98 | + | ||
| 99 | + **`realDtype`** (`string`): Actual data type. | ||
| 100 | + | ||
| 101 | + **`descName`** (`string`): Name description. | ||
| 102 | + | ||
| 103 | +- **at_npu::native::OpCommand& at_npu::native::OpCommand::Input(const c10::ArrayRef& dimListRef, at::IntArrayRef realShape, at::ScalarType toType = at::kDouble, CompileType compileType = CompileType::MEMORY_HOST_COMPILE_DEPENDENT, const string& realDtype = "")** | ||
| 104 | + | ||
| 105 | + Floating-point array input for `OpCommand`, typically provided on the CPU side, which is copied through H2D during execution. The return value type is `OpCommand`. | ||
| 106 | + | ||
| 107 | + **`dimListRef`** (`ArrayRef`): Input floating-point array. | ||
| 108 | + | ||
| 109 | + **`realShape`** (`IntArrayRef`): Input shape dimensions. | ||
| 110 | + | ||
| 111 | + **`toType`** (`ScalarType`): Target data type. | ||
| 112 | + | ||
| 113 | + **`compileType`** (`CompileType`): Compilation type. | ||
| 114 | + | ||
| 115 | + **`realDtype`** (`string`): Actual data type. | ||
| 116 | + | ||
| 117 | +- **at_npu::native::OpCommand& at_npu::native::OpCommand::Input(const c10::Scalar& input, const at::ScalarType type, CompileType compileType = CompileType::MEMORY_HOST_COMPILE_INDEPENDENT)** | ||
| 118 | + | ||
| 119 | + Scalar input for `OpCommand`, which is copied via H2D during execution. The return value type is `OpCommand`. | ||
| 120 | + | ||
| 121 | + **`input`** (`Scalar`): Input scalar reference. | ||
| 122 | + | ||
| 123 | + **`type`** (`ScalarType`): Target data type. | ||
| 124 | + | ||
| 125 | + **`compileType`** (`CompileType`): Compilation type. | ||
| 126 | + | ||
| 127 | +- **at_npu::native::OpCommand& at_npu::native::OpCommand::Inputs(const at::TensorList& inputs)** | ||
| 128 | + | ||
| 129 | + Tensor list input for `OpCommand`. The return value type is `OpCommand`. | ||
| 130 | + | ||
| 131 | + **`inputs`** (`TensorList`): Input tensor list reference. | ||
| 132 | + | ||
| 133 | +- **at_npu::native::OpCommand& at_npu::native::OpCommand::InputScalarToNPUTensor(const c10::Scalar& input, const at::ScalarType type)** | ||
| 134 | + | ||
| 135 | + Scalar input for `OpCommand`. The return value type is `OpCommand`. | ||
| 136 | + | ||
| 137 | + **`input`** (`Scalar`): Input scalar reference. | ||
| 138 | + | ||
| 139 | + **`type`** (`ScalarType`): Target data type. | ||
| 140 | + | ||
| 141 | +- **at_npu::native::OpCommand& at_npu::native::OpCommand::Output(at::Tensor& output, const string& descName = "", const c10::optional& sensitive_format = c10::nullopt, const string& realType = "")** | ||
| 142 | + | ||
| 143 | + Output tensor for `OpCommand`. The return value type is `OpCommand`. | ||
| 144 | + | ||
| 145 | + **`output`** (`Tensor`): Output tensor reference. | ||
| 146 | + | ||
| 147 | + **`descName`** (`string`): Name description. | ||
| 148 | + | ||
| 149 | + **`sensitive_format`** (`aclFormat`): Specific format requirements. | ||
| 150 | + | ||
| 151 | + **`realType`** (`string`): Actual data type. | ||
| 152 | + | ||
| 153 | +- **template at_npu::native::OpCommand& at_npu::native::OpCommand::Attr(const string& name, dataType value)** | ||
| 154 | + | ||
| 155 | + Sets an attribute for `OpCommand`. The return value type is `OpCommand`. | ||
| 156 | + | ||
| 157 | + **`name`** (`string`): Attribute name constant reference. | ||
| 158 | + | ||
| 159 | + **`value`** (`dataType`): Attribute value. | ||
| 160 | + | ||
| 161 | +- **template at_npu::native::OpCommand& at_npu::native::OpCommand::Attr(const string& name, dataType value, bool cond)** | ||
| 162 | + | ||
| 163 | + Conditionally sets an attribute for `OpCommand`. The return value type is `OpCommand`. | ||
| 164 | + | ||
| 165 | + **`name`** (`string`): Attribute name constant reference. | ||
| 166 | + | ||
| 167 | + **`value`** (`dataType`): Attribute value. | ||
| 168 | + | ||
| 169 | + **`cond`** (`bool`): Condition used to determine whether the attribute is set. If `False`, the attribute is not set. | ||
| 170 | + | ||
| 171 | +- **void at_npu::native::OpCommand::Run()** | ||
| 172 | + | ||
| 173 | + Executes the `OpCommand` operator. | ||
| 174 | + | ||
| 175 | +- **at_npu::native::OpCommand& at_npu::native::OpCommand::Sync(c10::SmallVector<int64_t, N\>& sync_index)** | ||
| 176 | + | ||
| 177 | + Sets synchronization indices for `OpCommand`. The return value type is `OpCommand`. | ||
| 178 | + | ||
| 179 | + **`sync_index`** (`SmallVector<int64_t, N>`): Reference to the indices to be synchronized. | ||
| 180 | + | ||
| 181 | +- **at_npu::native::OpCommand& at_npu::native::OpCommand::Sync()** | ||
| 182 | + | ||
| 183 | + Synchronizes `OpCommand` and blocks execution until the execution stream completes. | ||
| 184 | + | ||
| 185 | +## Supported Products | ||
| 186 | + | ||
| 187 | +- <term>Atlas training products</term> | ||
| 188 | +- <term>Atlas A2 training products</term> | ||
| 189 | +- <term>Atlas A3 training products</term> | ||
| 190 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,116 @@ | |||
| 1 | +# (beta) class c10_npu::NPUStream | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\core\npu\NPUStream.h | ||
| 6 | + | ||
| 7 | +## Function | ||
| 8 | + | ||
| 9 | +Implements NPU stream management functionality as an NPU stream class. An NPU stream is a linear execution sequence belonging to an NPU device. | ||
| 10 | + | ||
| 11 | +## Member Functions | ||
| 12 | + | ||
| 13 | +- **c10_npu::NPUStream::NPUStream(c10::Stream stream)** | ||
| 14 | + | ||
| 15 | + `NPUStream` constructor, which constructs an NPU stream from a specified stream. This function is identical to `c10::cuda::CUDAStream::CUDAStream(c10::Stream stream)`. | ||
| 16 | + | ||
| 17 | + **`stream`** (`Stream`): Specified input stream, which must be an NPU stream. | ||
| 18 | + | ||
| 19 | +- **c10_npu::NPUStream::NPUStream(Unchecked, c10::Stream stream)** | ||
| 20 | + | ||
| 21 | + `NPUStream` constructor, which constructs an NPU stream from a specified stream without checking whether it is an NPU stream. This function is identical to `c10::cuda::CUDAStream::CUDAStream(Unchecked, c10::Stream stream)`. | ||
| 22 | + | ||
| 23 | + **`stream`** (`Stream`): Specified input stream. | ||
| 24 | + | ||
| 25 | +- **c10_npu::NPUStream::\~NPUStream()** | ||
| 26 | + | ||
| 27 | + `NPUStream` destructor, which is identical to `c10::cuda::CUDAStream::~CUDAStream()`. | ||
| 28 | + | ||
| 29 | +- **bool c10_npu::NPUStream::operator==(const c10_npu::NPUStream& other)** | ||
| 30 | + | ||
| 31 | + Overloads the `==` operator for `NPUStream`. Returns `True` if the two streams are equal after conversion to `c10::Stream`. This function is identical to `bool c10::cuda::CUDAStream::operator==(const c10::cuda::CUDAStream& other)`. | ||
| 32 | + | ||
| 33 | + **`other`** (`NPUStream`): NPU stream to be compared. | ||
| 34 | + | ||
| 35 | +- **bool c10_npu::NPUStream::operator!=(const c10_npu::NPUStream& other)** | ||
| 36 | + | ||
| 37 | + Overloads the `!=` operator for `NPUStream`. Returns `True` if the two streams are not equal after conversion to `c10::Stream`. This function is identical to `bool c10::cuda::CUDAStream::operator!=(const c10::cuda::CUDAStream& other)`. | ||
| 38 | + | ||
| 39 | + **`other`** (`NPUStream`): NPU stream to be compared. | ||
| 40 | + | ||
| 41 | +- **c10_npu::NPUStream::operator aclrtStream()** | ||
| 42 | + | ||
| 43 | + Performs `aclrtStream` type conversion for `NPUStream`. This function is identical to `c10::cuda::CUDAStream::operator cudaStream_t()`. | ||
| 44 | + | ||
| 45 | +- **c10_npu::NPUStream::operator c10::Stream()** | ||
| 46 | + | ||
| 47 | + Performs `c10::Stream` type conversion for `NPUStream`. This function is identical to `c10::cuda::CUDAStream::operator c10::Stream()`. | ||
| 48 | + | ||
| 49 | +- **c10::DeviceType c10_npu::NPUStream::device_type()** | ||
| 50 | + | ||
| 51 | + Obtains the device type for `NPUStream`. The return type is `DeviceType`. This function is identical to `c10::DeviceType c10::cuda::CUDAStream::device_type()`. | ||
| 52 | + | ||
| 53 | +- **c10::DeviceIndex c10_npu::NPUStream::device_index()** | ||
| 54 | + | ||
| 55 | + Obtains the device ID for `NPUStream`. The return type is `DeviceIndex`. This function is identical to `c10::DeviceIndex c10::cuda::CUDAStream::device_index()`. | ||
| 56 | + | ||
| 57 | +- **c10::Device c10_npu::NPUStream::device()** | ||
| 58 | + | ||
| 59 | + Obtains the device for `NPUStream`. The return type is `Device`, which is guaranteed to represent an NPU device. This function is identical to `c10::Device c10::cuda::CUDAStream::device()`. | ||
| 60 | + | ||
| 61 | +- **c10::StreamId c10_npu::NPUStream::id()** | ||
| 62 | + | ||
| 63 | + Obtains the stream ID for `NPUStream`. The return type is `StreamId`. This function is identical to `c10::StreamId c10::cuda::CUDAStream::id()`. | ||
| 64 | + | ||
| 65 | +- **bool c10_npu::NPUStream::query()** | ||
| 66 | + | ||
| 67 | + Queries whether the stream has completed for `NPUStream`. The return type is `bool`. A return value of `True` indicates that all submitted work has completed. This function is identical to `bool c10::cuda::CUDAStream::query()`. | ||
| 68 | + | ||
| 69 | +- **void c10_npu::NPUStream::synchronize()** | ||
| 70 | + | ||
| 71 | + Synchronizes the stream for `NPUStream`, which blocks execution until all submitted work completes. This function is identical to `void c10::cuda::CUDAStream::synchronize()`. | ||
| 72 | + | ||
| 73 | +- **aclrtStream c10_npu::NPUStream::stream()** | ||
| 74 | + | ||
| 75 | + Queries the `aclrtStream` stream for `NPUStream`. The return type is `aclrtStream`. This function is identical to `cudaStream_t c10::cuda::CUDAStream::stream()`. | ||
| 76 | + | ||
| 77 | +- **c10::Stream c10_npu::NPUStream::unwrap()** | ||
| 78 | + | ||
| 79 | + Queries the `Stream` stream for `NPUStream`. The return type is `Stream`. This function is identical to `c10::Stream c10::cuda::CUDAStream::unwrap()`. | ||
| 80 | + | ||
| 81 | +- **struct c10::StreamData3 c10_npu::NPUStream::pack3()** | ||
| 82 | + | ||
| 83 | + Packs the stream for `NPUStream`. The return type is `StreamData3`. This function is identical to `struct c10::StreamData3 c10::cuda::CUDAStream::pack3()`. | ||
| 84 | + | ||
| 85 | +- **c10_npu::NPUStream c10_npu::NPUStream::unpack3(c10::StreamId stream_id, c10::DeviceIndex device_index, c10::DeviceType device_type)** | ||
| 86 | + | ||
| 87 | + Unpacks a stream from a `StreamData3` structure for `NPUStream`. The return type is `NPUStream`. This function is identical to `c10::cuda::CUDAStream c10::cuda::CUDAStream::unpack3(c10::StreamId stream_id, c10::DeviceIndex device_index, c10::DeviceType device_type)`. | ||
| 88 | + | ||
| 89 | + **`stream_id`** (`StreamId`): Stream ID stored in the `StreamData3` structure. | ||
| 90 | + | ||
| 91 | + **`device_index`** (`DeviceIndex`): Device ID stored in the `StreamData3` structure. | ||
| 92 | + | ||
| 93 | + **`device_type`** (`DeviceType`): Device type stored in the `StreamData3` structure. | ||
| 94 | + | ||
| 95 | +- **void c10_npu::NPUStream::setDataPreprocessStream(bool is_data_preprocess_stream)** | ||
| 96 | + | ||
| 97 | + Sets `NPUStream` as a data preprocessing stream. | ||
| 98 | + | ||
| 99 | + **`is_data_preprocess_stream`** (`bool`): `True` specifies the stream as a data preprocessing stream. | ||
| 100 | + | ||
| 101 | +- **bool c10_npu::NPUStream::isDataPreprocessStream()** | ||
| 102 | + | ||
| 103 | + Queries whether the stream is a data preprocessing stream for `NPUStream`. The return type is `bool`. A return value of `True` indicates that it is a data preprocessing stream. | ||
| 104 | + | ||
| 105 | +- **aclrtStream c10_npu::NPUStream::stream(const bool need_empty)** | ||
| 106 | + | ||
| 107 | + Queries the `aclrtStream` stream for `NPUStream`, allowing the input parameter `need_empty`. The return type is `aclrtStream`. | ||
| 108 | + | ||
| 109 | + **`need_empty`** (`bool`): `False` specifies to return the current stream directly instead of an empty stream. | ||
| 110 | + | ||
| 111 | +## Supported Products | ||
| 112 | + | ||
| 113 | +- <term>Atlas training products</term> | ||
| 114 | +- <term>Atlas A2 training products</term> | ||
| 115 | +- <term>Atlas A3 training products</term> | ||
| 116 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,94 @@ | |||
| 1 | +# (beta) struct c10_npu::NPUEvent | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\core\npu\NPUEvent.h | ||
| 6 | + | ||
| 7 | +## Function | ||
| 8 | + | ||
| 9 | +Implements NPU event management functionality as an event class. It can be used to monitor device progress, accurately measure elapsed time, or synchronize NPU streams. | ||
| 10 | + | ||
| 11 | +## Member Functions | ||
| 12 | + | ||
| 13 | +- **c10_npu::NPUEvent::NPUEvent()** | ||
| 14 | + | ||
| 15 | + Default constructor for `NPUEvent`. This function is identical to `at::cuda::CUDAEvent::CUDAEvent()`. | ||
| 16 | + | ||
| 17 | +- **c10_npu::NPUEvent::NPUEvent(unsigned int flags)** | ||
| 18 | + | ||
| 19 | + Constructs an `NPUEvent` with flags. This function is identical to `at::cuda::CUDAEvent::CUDAEvent(unsigned int flags)`. | ||
| 20 | + | ||
| 21 | + **`flags`** (`unsigned int`): Event type to be constructed. | ||
| 22 | + | ||
| 23 | +- **c10_npu::NPUEvent::\~NPUEvent()** | ||
| 24 | + | ||
| 25 | + Destructor for `NPUEvent`. This function is identical to `at::cuda::CUDAEvent::~CUDAEvent()`. | ||
| 26 | + | ||
| 27 | +- **c10_npu::NPUEvent::NPUEvent(c10_npu::NPUEvent&& other)** | ||
| 28 | + | ||
| 29 | + Move constructor for `NPUEvent`. This function is identical to `at::cuda::CUDAEvent::CUDAEvent(at::cuda::CUDAEvent&& other)`. | ||
| 30 | + | ||
| 31 | + `other` (`NPUEvent`): Object used to move-construct a new `NPUEvent` object. | ||
| 32 | + | ||
| 33 | +- **c10_npu::NPUEvent::operator aclrtEvent()** | ||
| 34 | + | ||
| 35 | + Performs type conversion from `NPUEvent` to `aclrtEvent`. This function is identical to `at::cuda::CUDAEvent::operator cudaEvent_t()`. | ||
| 36 | + | ||
| 37 | +- **c10::optional\<at::Device> c10_npu::NPUEvent::device()** | ||
| 38 | + | ||
| 39 | + Obtains the device type for `NPUEvent`. This function is identical to `c10::optional<at::Device> at::cuda::CUDAEvent::device()`. | ||
| 40 | + | ||
| 41 | +- **bool c10_npu::NPUEvent::isCreated()** | ||
| 42 | + | ||
| 43 | + Queries whether the event is created for `NPUEvent`. The return type is `bool`. Valid values are `True` (the event is created) or `False` (the event is not created). This function is identical to `bool at::cuda::CUDAEvent::isCreated()`. | ||
| 44 | + | ||
| 45 | +- **c10::DeviceIndex c10_npu::NPUEvent::device_index()** | ||
| 46 | + | ||
| 47 | + Obtains the device ID for `NPUEvent`. This function is identical to `c10::DeviceIndex at::cuda::CUDAEvent::device_index()`. | ||
| 48 | + | ||
| 49 | +- **aclrtEvent c10_npu::NPUEvent::event()** | ||
| 50 | + | ||
| 51 | + Queries the underlying ACL event for `NPUEvent`. The return type is `aclrtEvent`. This function is identical to `cudaEvent_t at::cuda::CUDAEvent::event()`. | ||
| 52 | + | ||
| 53 | +- **bool c10_npu::NPUEvent::query()** | ||
| 54 | + | ||
| 55 | + Queries whether the event is complete for `NPUEvent`. The return type is `bool`. Valid values are `True` (all submitted work is complete) or `False` (the work is not complete). This function is identical to `bool at::cuda::CUDAEvent::query()`. | ||
| 56 | + | ||
| 57 | +- **void c10_npu::NPUEvent::record()** | ||
| 58 | + | ||
| 59 | + Records the event for `NPUEvent`, which records the event on the current stream. This function is identical to `void at::cuda::CUDAEvent::record()`. | ||
| 60 | + | ||
| 61 | +- **void c10_npu::NPUEvent::record(const c10_npu::NPUStream& stream)** | ||
| 62 | + | ||
| 63 | + Records the event for `NPUEvent`, which records the event on a specified stream. This function is identical to `void at::cuda::CUDAEvent::record(const c10::cuda::CUDAStream& stream)`. | ||
| 64 | + | ||
| 65 | + **`stream`** (`NPUStream`): Stream on which the event is recorded. | ||
| 66 | + | ||
| 67 | +- **void c10_npu::NPUEvent::recordOnce(const c10_npu::NPUStream& stream)** | ||
| 68 | + | ||
| 69 | + Records the event for `NPUEvent`. Given a specified stream, the event is recorded once if it has not been recorded previously. This function is identical to `void at::cuda::CUDAEvent::recordOnce(const c10::cuda::CUDAStream& stream)`. | ||
| 70 | + | ||
| 71 | + **`stream`** (`NPUStream`): Stream on which the event is recorded. | ||
| 72 | + | ||
| 73 | +- **void c10_npu::NPUEvent::block(const c10_npu::NPUStream& stream)** | ||
| 74 | + | ||
| 75 | + Blocks a specified event for `NPUEvent`. This function is identical to `void at::cuda::CUDAEvent::block(const c10::cuda::CUDAStream& stream)`. | ||
| 76 | + | ||
| 77 | + **`stream`** (`NPUStream`): Stream to block. | ||
| 78 | + | ||
| 79 | +- **float c10_npu::NPUEvent::elapsed_time(const c10_npu::NPUEvent& other)** | ||
| 80 | + | ||
| 81 | + Queries the elapsed time between `NPUEvent` events. The return type is `float`, which represents the elapsed time in milliseconds from the recorded event to the current event. This function is identical to `float at::cuda::CUDAEvent::elapsed_time(const at::cuda::CUDAEvent& other)`. | ||
| 82 | + | ||
| 83 | + **`other`** (`NPUEvent`): Destination event used to compute the elapsed time. | ||
| 84 | + | ||
| 85 | +- **void c10_npu::NPUEvent::synchronize()** | ||
| 86 | + | ||
| 87 | + Synchronizes the event for `NPUEvent`, which blocks execution until the event completes execution. This function is identical to `void at::cuda::CUDAEvent::synchronize()`. | ||
| 88 | + | ||
| 89 | +## Supported Products | ||
| 90 | + | ||
| 91 | +- <term>Atlas training products</term> | ||
| 92 | +- <term>Atlas A2 training products</term> | ||
| 93 | +- <term>Atlas A3 training products</term> | ||
| 94 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,16 @@ | |||
| 1 | +# (beta) struct c10_npu::NPUHooksArgs | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\core\npu\NPUHooksInterface.h | ||
| 6 | + | ||
| 7 | +## Function | ||
| 8 | + | ||
| 9 | +Provides parameters for NPU hooks as a hook parameter class. | ||
| 10 | + | ||
| 11 | +## Supported Products | ||
| 12 | + | ||
| 13 | +- <term>Atlas training products</term> | ||
| 14 | +- <term>Atlas A2 training products</term> | ||
| 15 | +- <term>Atlas A3 training products</term> | ||
| 16 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,24 @@ | |||
| 1 | +# (beta) struct c10_npu::NPUHooksInterface | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\core\npu\NPUHooksInterface.h | ||
| 6 | + | ||
| 7 | +## Function | ||
| 8 | + | ||
| 9 | +Provides NPU hook APIs as a hook interface class. | ||
| 10 | + | ||
| 11 | +## Member Functions | ||
| 12 | + | ||
| 13 | +- **const at::Generator& c10_npu::NPUHooksInterface::getDefaultGenerator(c10::DeviceIndex device_index)** | ||
| 14 | + | ||
| 15 | +Obtains the default random number generator for `NPUHooksInterface`. This function is identical to `const at::Generator& at::CUDAHooksInterface::getDefaultCUDAGenerator(c10::DeviceIndex device_index = -1)`. | ||
| 16 | + | ||
| 17 | +**`device_index`** (`DeviceIndex`): NPU device ID. | ||
| 18 | + | ||
| 19 | +## Supported Products | ||
| 20 | + | ||
| 21 | +- <term>Atlas training products</term> | ||
| 22 | +- <term>Atlas A2 training products</term> | ||
| 23 | +- <term>Atlas A3 training products</term> | ||
| 24 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,26 @@ | |||
| 1 | +# (beta) torch::npu::synchronize | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\libs\init_npu.h | ||
| 6 | + | ||
| 7 | +## Prototype | ||
| 8 | + | ||
| 9 | +```cpp | ||
| 10 | +void torch::npu::synchronize(int64_t device_index = -1) | ||
| 11 | +``` | ||
| 12 | + | ||
| 13 | +## Function | ||
| 14 | + | ||
| 15 | +Synchronizes the NPU device. This function blocks the current thread until all computation tasks submitted to the NPU device complete execution. This function is identical to `void torch::cuda::synchronize(int64_t device_index = -1)`. | ||
| 16 | + | ||
| 17 | +## Parameters | ||
| 18 | + | ||
| 19 | +**`device_index`** (`int64_t`): Optional. Device index used for synchronization. The default value is `-1`, which specifying to synchronize the current device. | ||
| 20 | + | ||
| 21 | +## Supported Products | ||
| 22 | + | ||
| 23 | +- <term>Atlas training products</term> | ||
| 24 | +- <term>Atlas A2 training products</term> | ||
| 25 | +- <term>Atlas A3 training products</term> | ||
| 26 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,22 @@ | |||
| 1 | +# (beta) torch_npu::finalize_npu | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\libs\init_npu.h | ||
| 6 | + | ||
| 7 | +## Prototype | ||
| 8 | + | ||
| 9 | +```cpp | ||
| 10 | +void torch_npu::finalize_npu() | ||
| 11 | +``` | ||
| 12 | + | ||
| 13 | +## Function | ||
| 14 | + | ||
| 15 | +Deinitializes the NPU device, which releases NPU resources. | ||
| 16 | + | ||
| 17 | +## Supported Products | ||
| 18 | + | ||
| 19 | +- <term>Atlas training products</term> | ||
| 20 | +- <term>Atlas A2 training products</term> | ||
| 21 | +- <term>Atlas A3 training products</term> | ||
| 22 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1,30 @@ | |||
| 1 | +# (beta) torch_npu::init_npu | ||
| 2 | + | ||
| 3 | +## Definition File | ||
| 4 | + | ||
| 5 | +torch_npu\csrc\libs\init_npu.h | ||
| 6 | + | ||
| 7 | +## Prototype | ||
| 8 | + | ||
| 9 | +```cpp | ||
| 10 | +void torch_npu::init_npu(const c10::DeviceIndex device_index = 0) | ||
| 11 | +void torch_npu::init_npu(const std::string& device_str) | ||
| 12 | +void torch_npu::init_npu(const at::Device& device) | ||
| 13 | +``` | ||
| 14 | + | ||
| 15 | +## Function | ||
| 16 | + | ||
| 17 | +Initializes the NPU device. | ||
| 18 | + | ||
| 19 | +## Parameters | ||
| 20 | + | ||
| 21 | +- **`device_index`** (`DeviceIndex`): ID of the NPU device to be initialized. The default value is `0`. | ||
| 22 | +- **`device_str`** (`string`): Name of the device to be initialized. | ||
| 23 | +- **`device`** (`Device`): NPU device to be initialized. | ||
| 24 | + | ||
| 25 | +## Supported Products | ||
| 26 | + | ||
| 27 | +- <term>Atlas training products</term> | ||
| 28 | +- <term>Atlas A2 training products</term> | ||
| 29 | +- <term>Atlas A3 training products</term> | ||
| 30 | +- <term>Atlas inference products</term> | ||
| @@ -0,0 +1 @@ | |||
| 1 | +# C++ APIs | ||
| @@ -0,0 +1,240 @@ | |||
| 1 | +# C++ APIs | ||
| 2 | + | ||
| 3 | +This section describes the extended C++ APIs provided by TorchNPU for deep optimization. | ||
| 4 | + | ||
| 5 | +## Standard C++ APIs | ||
| 6 | + | ||
| 7 | +The standard C++ APIs provide fundamental capabilities such as NPU device initialization and release, device and stream management, event and random number management, tensor format handling, and distributed communication. For details, see Table 1. The symbols of these APIs are uniformly exported by `libtorch_npu.so`. | ||
| 8 | + | ||
| 9 | +**Table 1** C++ APIs | ||
| 10 | +<a name="table1728514451512"></a> | ||
| 11 | +<table><thead align="left"><tr id="row1128511418151"><th class="cellrowborder" valign="top" width="36.13%" id="mcps1.2.3.1.1"><p id="p102854481517"><a name="p102854481517"></a><a name="p102854481517"></a>API</p> | ||
| 12 | +</th> | ||
| 13 | +<th class="cellrowborder" valign="top" width="63.870000000000005%" id="mcps1.2.3.1.2"><p id="p102851481518"><a name="p102851481518"></a><a name="p102851481518"></a>Description</p> | ||
| 14 | +</th> | ||
| 15 | +</tr> | ||
| 16 | +</thead> | ||
| 17 | +<tbody><tr id="row182857401519"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p14285194191518"><a name="p14285194191518"></a><a name="p14285194191518"></a><a href="(beta)torch_npu-init_npu.md">(beta) torch_npu::init_npu</a></p> | ||
| 18 | +</td> | ||
| 19 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p3285204101519"><a name="p3285204101519"></a><a name="p3285204101519"></a>Initializes the NPU device.</p> | ||
| 20 | +</td> | ||
| 21 | +</tr> | ||
| 22 | +<tr id="row374312587556"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p11667145619"><a name="p11667145619"></a><a name="p11667145619"></a><a href="(beta)torch_npu-finalize_npu.md">(beta) torch_npu::finalize_npu</a></p> | ||
| 23 | +</td> | ||
| 24 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p1549042791610"><a name="p1549042791610"></a><a name="p1549042791610"></a>Deinitializes the NPU device, which releases NPU resources.</p> | ||
| 25 | +</td> | ||
| 26 | +</tr> | ||
| 27 | +<tr id="row18285164121511"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p52853416159"><a name="p52853416159"></a><a name="p52853416159"></a><a href="(beta)torch-npu-synchronize.md">(beta) torch::npu::synchronize</a></p> | ||
| 28 | +</td> | ||
| 29 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p14285184141513"><a name="p14285184141513"></a><a name="p14285184141513"></a>Synchronizes the NPU device. This function is identical to <code>void torch::cuda::synchronize(int64_t <em id="i5516134825616"><a name="i5516134825616"></a><a name="i5516134825616"></a>device_index </em>= -1)</code>.</p> | ||
| 30 | +</td> | ||
| 31 | +</tr> | ||
| 32 | +<tr id="row13285194191516"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p13285154101510"><a name="p13285154101510"></a><a name="p13285154101510"></a><a href="(beta)c10-npu-current_device.md">(beta) c10::npu::current_device</a></p> | ||
| 33 | +</td> | ||
| 34 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p42853421514"><a name="p42853421514"></a><a name="p42853421514"></a>Obtains the NPU device ID. The return value type is <code>DeviceIndex</code>. This function is identical to <code>c10::DeviceIndex c10::cuda::current_device()</code>.</p> | ||
| 35 | +</td> | ||
| 36 | +</tr> | ||
| 37 | +<tr id="row7285114101517"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p12850416158"><a name="p12850416158"></a><a name="p12850416158"></a><a href="(beta)at-Device.md">(beta) at::Device</a></p> | ||
| 38 | +</td> | ||
| 39 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p6285046154"><a name="p6285046154"></a><a name="p6285046154"></a>After TorchNPU is installed, the <code>NPU</code> field is supported for the Device type, allowing a device to be specified using a string description.</p> | ||
| 40 | +</td> | ||
| 41 | +</tr> | ||
| 42 | +<tr id="row1482153472113"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p197043314361"><a name="p197043314361"></a><a name="p197043314361"></a><a href="(beta)struct-c10_npu-NPUEvent.md">(beta) struct c10_npu::NPUEvent</a></p> | ||
| 43 | +</td> | ||
| 44 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p1297013314368"><a name="p1297013314368"></a><a name="p1297013314368"></a>Implements NPU event management functionality as an event class. It can be used to monitor device progress, accurately measure elapsed time, or synchronize NPU streams.</p> | ||
| 45 | +</td> | ||
| 46 | +</tr> | ||
| 47 | +<tr id="row134831934102116"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p16592356133615"><a name="p16592356133615"></a><a name="p16592356133615"></a><a href="(beta)class-at_npu-NPUGeneratorImpl.md">(beta) class at_npu::NPUGeneratorImpl</a></p> | ||
| 48 | +</td> | ||
| 49 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p142133114239"><a name="p142133114239"></a><a name="p142133114239"></a>Acts as a random number generator class that implements NPU device random number functionality and supports various random-dependent methods.</p> | ||
| 50 | +</td> | ||
| 51 | +</tr> | ||
| 52 | +<tr id="row8483934102114"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p6592656183613"><a name="p6592656183613"></a><a name="p6592656183613"></a><a href="(beta)at_npu-detail-getDefaultNPUGenerator.md">(beta) at_npu::detail::getDefaultNPUGenerator</a></p> | ||
| 53 | +</td> | ||
| 54 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p1592956123617"><a name="p1592956123617"></a><a name="p1592956123617"></a>Obtains the default generator for an NPU device. The return value type is <code>Generator</code>, which is identical to <code>at::Generator& at::cuda::detail::getDefaultCUDAGenerator(c10::DeviceIndex <em id="i1647844813422"><a name="i1647844813422"></a><a name="i1647844813422"></a>device_index</em> = -1)</code>.</p> | ||
| 55 | +</td> | ||
| 56 | +</tr> | ||
| 57 | +<tr id="row124831534202119"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p15593155612361"><a name="p15593155612361"></a><a name="p15593155612361"></a><a href="(beta)at_npu-detail-createNPUGenerator.md">(beta) at_npu::detail::createNPUGenerator</a></p> | ||
| 58 | +</td> | ||
| 59 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p1059375683612"><a name="p1059375683612"></a><a name="p1059375683612"></a>Creates the default generator for an NPU device. The return value type is <code>Generator</code>, which is identical to <code>at::Generator at::cuda::detail::createCUDAGenerator(c10::DeviceIndex <em>device_index</em> = -1)</code>.</p> | ||
| 60 | +</td> | ||
| 61 | +</tr> | ||
| 62 | +<tr id="row1582212372214"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p1145460103719"><a name="p1145460103719"></a><a name="p1145460103719"></a><a href="(beta)class-c10_npu-NPUStream.md">(beta) class c10_npu::NPUStream</a></p> | ||
| 63 | +</td> | ||
| 64 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p1245413083714"><a name="p1245413083714"></a><a name="p1245413083714"></a>Implements NPU stream management functionality as an NPU stream class. An NPU stream is a linear execution sequence belonging to an NPU device.</p> | ||
| 65 | +</td> | ||
| 66 | +</tr> | ||
| 67 | +<tr id="row78220372215"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p174555033712"><a name="p174555033712"></a><a name="p174555033712"></a><a href="(beta)c10_npu-getNPUStreamFromPool.md">(beta) c10_npu::getNPUStreamFromPool</a></p> | ||
| 68 | +</td> | ||
| 69 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p124552011374"><a name="p124552011374"></a><a name="p124552011374"></a>Obtains a new stream from the NPU stream pool. The return value type is <code>NPUStream</code>. This function is identical to <code>c10::cuda::CUDAStream c10::cuda::getStreamFromPool(const bool <em id="i117561711437"><a name="i117561711437"></a><a name="i117561711437"></a>isHighPriority</em> = false, c10::DeviceIndex <em id="i66333254311"><a name="i66333254311"></a><a name="i66333254311"></a>device</em> = -1)</code>.</p> | ||
| 70 | +</td> | ||
| 71 | +</tr> | ||
| 72 | +<tr id="row11822183742111"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p145511016377"><a name="p145511016377"></a><a name="p145511016377"></a><a href="(beta)c10_npu-getDefaultNPUStream.md">(beta) c10_npu::getDefaultNPUStream</a></p> | ||
| 73 | +</td> | ||
| 74 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p245530143713"><a name="p245530143713"></a><a name="p245530143713"></a>Obtains the default NPU stream. The return value type is <code>NPUStream</code>. Its function and usage are identical to <code>c10::cuda::CUDAStream c10::cuda::getDefaultCUDAStream(c10::DeviceIndex <em id="i1692521216432"><a name="i1692521216432"></a><a name="i1692521216432"></a>device_index</em> = -1)</code>.</p> | ||
| 75 | +</td> | ||
| 76 | +</tr> | ||
| 77 | +<tr id="row6822103772114"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p2455402378"><a name="p2455402378"></a><a name="p2455402378"></a><a href="(beta)c10_npu-getCurrentNPUStream.md">(beta) c10_npu::getCurrentNPUStream</a></p> | ||
| 78 | +</td> | ||
| 79 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p104551505370"><a name="p104551505370"></a><a name="p104551505370"></a>Obtains the current NPU stream through a device ID. The return value type is <code>NPUStream</code>, which is identical to <code>c10::CUDA::CUDAStream c10::cuda::getCurrentCUDAStream(c10::DeviceIndex <em>device_index</em> = -1)</code>.</p> | ||
| 80 | +</td> | ||
| 81 | +</tr> | ||
| 82 | +<tr id="row1324644442114"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p10455150143713"><a name="p10455150143713"></a><a name="p10455150143713"></a><a href="(beta)c10_npu-setCurrentNPUStream.md">(beta) c10_npu::setCurrentNPUStream</a></p> | ||
| 83 | +</td> | ||
| 84 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p184550013377"><a name="p184550013377"></a><a name="p184550013377"></a>Sets the current NPU stream. This function is identical to <code>void c10::cuda::setCurrentCUDAStream(c10::cuda::CUDAStream <em id="i1428123154319"><a name="i1428123154319"></a><a name="i1428123154319"></a>stream</em>)</code>.</p> | ||
| 85 | +</td> | ||
| 86 | +</tr> | ||
| 87 | +<tr id="row11246164452110"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p1364718411376"><a name="p1364718411376"></a><a name="p1364718411376"></a><a href="(beta)class-at_npu-native-OpCommand.md">(beta) class at_npu::native::OpCommand</a></p> | ||
| 88 | +</td> | ||
| 89 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p11647446372"><a name="p11647446372"></a><a name="p11647446372"></a>Acts as a class encapsulating low-level operator calls and implements underlying operator execution on NPU devices.</p> | ||
| 90 | +</td> | ||
| 91 | +</tr> | ||
| 92 | +<tr id="row20187152616544"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p1935915613234"><a name="p1935915613234"></a><a name="p1935915613234"></a><a href="(beta)struct-c10_npu-NPUHooksInterface.md">(beta) struct c10_npu::NPUHooksInterface</a></p> | ||
| 93 | +</td> | ||
| 94 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p83135713258"><a name="p83135713258"></a><a name="p83135713258"></a>Provides NPU hook APIs as a hook interface class.</p> | ||
| 95 | +</td> | ||
| 96 | +</tr> | ||
| 97 | +<tr id="row518051813244"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p6180191892414"><a name="p6180191892414"></a><a name="p6180191892414"></a><a href="(beta)struct-c10_npu-NPUHooksArgs.md">(beta) struct c10_npu::NPUHooksArgs</a></p> | ||
| 98 | +</td> | ||
| 99 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p51808188246"><a name="p51808188246"></a><a name="p51808188246"></a>Provides parameters for NPU hooks as a hook parameter class.</p> | ||
| 100 | +</td> | ||
| 101 | +</tr> | ||
| 102 | +<tr id="row1728493216243"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p1328483211244"><a name="p1328483211244"></a><a name="p1328483211244"></a><a href="(beta)c10_npu-device_count.md">(beta) c10_npu::device_count</a></p> | ||
| 103 | +</td> | ||
| 104 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p1028463217241"><a name="p1028463217241"></a><a name="p1028463217241"></a>Obtains the number of NPU devices. The return value type is <code>DeviceIndex</code>. This function is identical to <code>c10::DeviceIndex c10::cuda::device_count()</code>.</p> | ||
| 105 | +</td> | ||
| 106 | +</tr> | ||
| 107 | +<tr id="row1848933022417"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p124901730162414"><a name="p124901730162414"></a><a name="p124901730162414"></a><a href="(beta)c10_npu-GetDevice.md">(beta) c10_npu::GetDevice</a></p> | ||
| 108 | +</td> | ||
| 109 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p2490183092411"><a name="p2490183092411"></a><a name="p2490183092411"></a>Obtains the NPU device ID. The return value type is <code>aclError</code>. This function is identical to <code>cudaError_t c10::cuda::GetDevice(int *<em id="i5414123319436"><a name="i5414123319436"></a><a name="i5414123319436"></a>device</em>)</code>.</p> | ||
| 110 | +</td> | ||
| 111 | +</tr> | ||
| 112 | +<tr id="row017092012245"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p101705205249"><a name="p101705205249"></a><a name="p101705205249"></a><a href="(beta)c10_npu-SetDevice.md">(beta) c10_npu::SetDevice</a></p> | ||
| 113 | +</td> | ||
| 114 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p6171172042412"><a name="p6171172042412"></a><a name="p6171172042412"></a>Sets the NPU device. The return value type is <code>aclError</code>. This function is identical to <code>cudaError_t c10::cuda::SetDevice(int <em id="i32541536164317"><a name="i32541536164317"></a><a name="i32541536164317"></a>device</em>)</code>.</p> | ||
| 115 | +</td> | ||
| 116 | +</tr> | ||
| 117 | +<tr id="row166762910245"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p186712293248"><a name="p186712293248"></a><a name="p186712293248"></a><a href="(beta)c10_npu-current_device.md">(beta) c10_npu::current_device</a></p> | ||
| 118 | +</td> | ||
| 119 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p06702919248"><a name="p06702919248"></a><a name="p06702919248"></a>Obtains the NPU device ID. The return value type is <code>DeviceIndex</code>, which represents the obtained device ID. This function is identical to <code>c10::DeviceIndex c10::cuda::current_device()</code> in PyTorch 1.11.0. The main difference from <code>c10_npu::GetDevice</code> is that this function includes additional error checking.</p> | ||
| 120 | +</td> | ||
| 121 | +</tr> | ||
| 122 | +<tr id="row193943155249"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p93941215142414"><a name="p93941215142414"></a><a name="p93941215142414"></a><a href="(beta)c10_npu-set_device.md">(beta) c10_npu::set_device</a></p> | ||
| 123 | +</td> | ||
| 124 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p1539413151242"><a name="p1539413151242"></a><a name="p1539413151242"></a>Sets the NPU device. This function is identical to <code>void c10::cuda::set_device(c10::DeviceIndex <em id="i13672341174314"><a name="i13672341174314"></a><a name="i13672341174314"></a>device</em>)</code>. The main difference from <code>c10_npu::SetDevice</code> is that this function includes additional error checking.</p> | ||
| 125 | +</td> | ||
| 126 | +</tr> | ||
| 127 | +<tr id="row14242145315620"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p2024245305610"><a name="p2024245305610"></a><a name="p2024245305610"></a><a href="(beta)c10_npu-warning_state.md">(beta) c10_npu::warning_state</a></p> | ||
| 128 | +</td> | ||
| 129 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p142421953125611"><a name="p142421953125611"></a><a name="p142421953125611"></a>Obtains the current synchronization warning level. The return type is the <code>WarningState</code> enum class, which can be <code>L_DISABLED</code> (no warning), <code>L_WARN</code> (warning), or <code>L_ERROR</code> (error). This function is identical to <code>WarningState& c10::cuda::warning_state()</code>.</p> | ||
| 130 | +</td> | ||
| 131 | +</tr> | ||
| 132 | +<tr id="row0199201317247"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p719931392415"><a name="p719931392415"></a><a name="p719931392415"></a><a href="(beta)c10_npu-warn_or_error_on_sync.md">(beta) c10_npu::warn_or_error_on_sync</a></p> | ||
| 133 | +</td> | ||
| 134 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p4199313142416"><a name="p4199313142416"></a><a name="p4199313142416"></a>Raises a warning or error during NPU synchronization, with no return value. This function raises an error or logs a warning based on the current warning level, which is identical to <code>void c10::cuda::warn_or_error_on_sync()</code>.</p> | ||
| 135 | +</td> | ||
| 136 | +</tr> | ||
| 137 | +<tr id="row36118212415"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p76215262419"><a name="p76215262419"></a><a name="p76215262419"></a><a href="(beta)at_npu-native-get_npu_format.md">(beta) at_npu::native::get_npu_format</a></p> | ||
| 138 | +</td> | ||
| 139 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p49191458195"><a name="p49191458195"></a><a name="p49191458195"></a>Obtains the format information of an NPU tensor. The return value type is <code>int64_t</code>, which represents the obtained NPU tensor format data.</p> | ||
| 140 | +</td> | ||
| 141 | +</tr> | ||
| 142 | +<tr id="row395820517240"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p1495814522412"><a name="p1495814522412"></a><a name="p1495814522412"></a><a href="(beta)at_npu-native-get_npu_storage_sizes.md">(beta) at_npu::native::get_npu_storage_sizes</a></p> | ||
| 143 | +</td> | ||
| 144 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p17958115172419"><a name="p17958115172419"></a><a name="p17958115172419"></a>Obtains the memory size of an NPU tensor. The return value type is <code>vector<int64_t></code>, which represents the allocated memory size of the NPU tensor.</p> | ||
| 145 | +</td> | ||
| 146 | +</tr> | ||
| 147 | +<tr id="row978610882412"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p678619815247"><a name="p678619815247"></a><a name="p678619815247"></a><a href="(beta)at_npu-native-npu_format_cast.md">(beta) at_npu::native::npu_format_cast</a></p> | ||
| 148 | +</td> | ||
| 149 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p1478628132414"><a name="p1478628132414"></a><a name="p1478628132414"></a>Converts the format of an NPU tensor. The return value type is <code>Tensor</code>, which represents the converted tensor.</p> | ||
| 150 | +</td> | ||
| 151 | +</tr> | ||
| 152 | +<tr id="row178951326102411"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p98951326142411"><a name="p98951326142411"></a><a name="p98951326142411"></a><a href="(beta)at_npu-native-empty_with_format.md">(beta) at_npu::native::empty_with_format</a></p> | ||
| 153 | +</td> | ||
| 154 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p12895172619247"><a name="p12895172619247"></a><a name="p12895172619247"></a>Obtains an empty NPU tensor in a specified format. The return value type is <code>Tensor</code>, which represents the obtained empty tensor.</p> | ||
| 155 | +</td> | ||
| 156 | +</tr> | ||
| 157 | +<tr id="row15172411122412"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p1317211162415"><a name="p1317211162415"></a><a name="p1317211162415"></a><a href="(beta)c10_npu-c10_npu_get_error_message.md">(beta) c10_npu::c10_npu_get_error_message</a></p> | ||
| 158 | +</td> | ||
| 159 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p61725119244"><a name="p61725119244"></a><a name="p61725119244"></a>Obtains error messages. The return value type is <code>char *</code>, which represents the obtained error message string.</p> | ||
| 160 | +</td> | ||
| 161 | +</tr> | ||
| 162 | +<tr id="row55131018125916"><td class="cellrowborder" valign="top" width="36.13%" headers="mcps1.2.3.1.1 "><p id="p75141718135919"><a name="p75141718135919"></a><a name="p75141718135919"></a><a href="(beta)at_npu-native-npu_dropout_gen_mask.md">(beta) at_npu::native::npu_dropout_gen_mask</a></p> | ||
| 163 | +</td> | ||
| 164 | +<td class="cellrowborder" valign="top" width="63.870000000000005%" headers="mcps1.2.3.1.2 "><p id="p1032517615010"><a name="p1032517615010"></a><a name="p1032517615010"></a>Generates a random mask based on the probability <code>p</code> during training, which is used to set elements to zero.</p> | ||
| 165 | +</td> | ||
| 166 | +</tr><tr><td class="cellrowborder" valign="top" width="38.61%" headers="mcps1.2.3.1.1 "><p><a href="at_npu-native-empty_with_swapped_memory.md">at_npu.native-empty_with_swapped_memory</a></p> | ||
| 167 | +</td> | ||
| 168 | +<td class="cellrowborder" valign="top" width="61.39%" headers="mcps1.2.3.1.2 "><p>Allocates a tensor with its device type set to NPU, while its actual memory resides on the host side.</p> | ||
| 169 | +</td> | ||
| 170 | +</tr> | ||
| 171 | +<tr><td class="cellrowborder" valign="top" width="38.61%" headers="mcps1.2.3.1.1 "><p><a href="c10_npu-NPUStreamGuard.md">c10_npu::NPUStreamGuard</a></p> | ||
| 172 | +</td> | ||
| 173 | +<td class="cellrowborder" valign="top" width="61.39%" headers="mcps1.2.3.1.2 "><p>Manages the NPU device stream within its execution scope. This guard is identical to <code>c10::cuda::CUDAStreamGuard</code>.</p> | ||
| 174 | +</td> | ||
| 175 | +</tr> | ||
| 176 | +<tr><td class="cellrowborder" valign="top" width="38.61%" headers="mcps1.2.3.1.1 "><p><a href="c10_npu-NPUStreamGuard-current_device.md">c10_npu::NPUStreamGuard::current_device</a></p> | ||
| 177 | +</td> | ||
| 178 | +<td class="cellrowborder" valign="top" width="61.39%" headers="mcps1.2.3.1.2 "><p>Returns the current device set by the guard.</p> | ||
| 179 | +</td> | ||
| 180 | +</tr> | ||
| 181 | +<tr><td class="cellrowborder" valign="top" width="38.61%" headers="mcps1.2.3.1.1 "><p><a href="c10_npu-NPUStreamGuard-current_stream.md">c10_npu::NPUStreamGuard::current_stream</a></p> | ||
| 182 | +</td> | ||
| 183 | +<td class="cellrowborder" valign="top" width="61.39%" headers="mcps1.2.3.1.2 "><p>Returns the stream currently managed by the guard.</p> | ||
| 184 | +</td> | ||
| 185 | +</tr> | ||
| 186 | +<tr><td class="cellrowborder" valign="top" width="38.61%" headers="mcps1.2.3.1.1 "><p><a href="c10_npu-NPUStreamGuard-NPUStreamGuard.md">c10_npu::NPUStreamGuard::NPUStreamGuard</a></p> | ||
| 187 | +</td> | ||
| 188 | +<td class="cellrowborder" valign="top" width="61.39%" headers="mcps1.2.3.1.2 "><p>Creates a stream guard.</p> | ||
| 189 | +</td> | ||
| 190 | +</tr> | ||
| 191 | +<tr><td class="cellrowborder" valign="top" width="38.61%" headers="mcps1.2.3.1.1 "><p><a href="c10_npu-NPUStreamGuard-original_device.md">c10_npu::NPUStreamGuard::original_device</a></p> | ||
| 192 | +</td> | ||
| 193 | +<td class="cellrowborder" valign="top" width="61.39%" headers="mcps1.2.3.1.2 "><p>Returns the NPU device captured during guard construction.</p> | ||
| 194 | +</td> | ||
| 195 | +</tr> | ||
| 196 | +<tr><td class="cellrowborder" valign="top" width="38.61%" headers="mcps1.2.3.1.1 "><p><a href="c10_npu-NPUStreamGuard-original_stream.md">c10_npu::NPUStreamGuard::original_stream</a></p> | ||
| 197 | +</td> | ||
| 198 | +<td class="cellrowborder" valign="top" width="61.39%" headers="mcps1.2.3.1.2 "><p>Returns the NPU stream set during guard construction.</p> | ||
| 199 | +</td> | ||
| 200 | +</tr> | ||
| 201 | +<tr><td class="cellrowborder" valign="top" width="38.61%" headers="mcps1.2.3.1.1 "><p><a href="c10_npu-NPUStreamGuard-reset_stream.md">c10_npu::NPUStreamGuard::reset_stream</a></p> | ||
| 202 | +</td> | ||
| 203 | +<td class="cellrowborder" valign="top" width="61.39%" headers="mcps1.2.3.1.2 "><p>Resets the guard to a new NPU stream.</p> | ||
| 204 | +</td> | ||
| 205 | +</tr> | ||
| 206 | +<tr><td class="cellrowborder" valign="top" width="38.61%" headers="mcps1.2.3.1.1 "><p><a href="c10_npu-stream_synchronize.md">c10_npu::stream_synchronize</a></p> | ||
| 207 | +</td> | ||
| 208 | +<td class="cellrowborder" valign="top" width="61.39%" headers="mcps1.2.3.1.2 "><p>Synchronizes an NPU stream. This function is identical to <code>c10::cuda::stream_synchronize</code>.</p> | ||
| 209 | +</td> | ||
| 210 | +</tr> | ||
| 211 | +<tr><td class="cellrowborder" valign="top" width="38.61%" headers="mcps1.2.3.1.1 "><p><a href="c10d_npu-ProcessGroupHCCL.md">c10d_npu::ProcessGroupHCCL</a></p> | ||
| 212 | +</td> | ||
| 213 | +<td class="cellrowborder" valign="top" width="61.39%" headers="mcps1.2.3.1.2 "><p>Inherits from <code>c10d::Backend</code> and implements the <code>HCCL</code> backend APIs for communication operators.</p> | ||
| 214 | +</td> | ||
| 215 | +</tr> | ||
| 216 | +<tr><td class="cellrowborder" valign="top" width="38.61%" headers="mcps1.2.3.1.1 "><p><a href="c10d_npu-ProcessGroupHCCL-batch_isend_irecv.md">c10d_npu::ProcessGroupHCCL::batch_isend_irecv</a></p> | ||
| 217 | +</td> | ||
| 218 | +<td class="cellrowborder" valign="top" width="61.39%" headers="mcps1.2.3.1.2 "><p>Sends or receives a batch of tensors. Asynchronously processes each point-to-point (P2P) operation in the sequence. Returns the corresponding request handle.</p> | ||
| 219 | +</td> | ||
| 220 | +</tr> | ||
| 221 | +</tbody> | ||
| 222 | +</table> | ||
| 223 | + | ||
| 224 | +## LibTorch Stable ABI APIs | ||
| 225 | + | ||
| 226 | +The PyTorch community provides LibTorch Stable ABI capabilities, offering a set of stable binary APIs that allow custom operators and C++ extensions to maintain cross-version compatibility without being bound to a specific PyTorch version. TorchNPU supplements NPU-side capabilities and provides the following APIs: | ||
| 227 | + | ||
| 228 | +- [`torch_npu/csrc/inductor/aoti_torch/c/shim.h`](https://gitcode.com/Ascend/pytorch/blob/master/torch_npu/csrc/inductor/aoti_torch/c/shim.h): Supplements the NPU device type and NPU-related Stable ABI APIs. | ||
| 229 | +- [`torch_npu/csrc/inductor/aoti_torch/generated/c_shim_npu.h`](https://gitcode.com/Ascend/pytorch/blob/master/torch_npu/csrc/inductor/aoti_torch/generated/v2r13/c_shim_npu.h): Provides Stable ABI declarations for NPU operators, with API names following the pattern `aoti_torch_npu_*`. | ||
| 230 | +- `libtorch_npu.so`: Exports the symbols for the newly added NPU Stable ABI APIs. | ||
| 231 | + | ||
| 232 | +## ATB C++ API Overview | ||
| 233 | + | ||
| 234 | +TorchNPU provides a set of ATB C++ APIs to invoke exposed ATB capabilities from C++. This set of APIs is exposed through a unified entry-point header file. | ||
| 235 | + | ||
| 236 | +- [`op_plugin/include/atb_ops.h`](../../../../op_plugin/include/atb_ops.h): Unified entry-point header file for ATB C++ APIs. | ||
| 237 | +- [`op_plugin/config/atb_ops.yaml`](../../../../op_plugin/config/atb_ops.yaml): List of currently exposed APIs. | ||
| 238 | +- `libop_plugin_atb.so`: Exports the corresponding symbols for the ATB C++ APIs mentioned above. | ||
| 239 | + | ||
| 240 | +These APIs can be accessed through the `at_npu::native::atb` namespace. | ||
| @@ -0,0 +1,49 @@ | |||
| 1 | +# at_npu::native::empty_with_swapped_memory | ||
| 2 | + | ||
| 3 | +## Supported Products | ||
| 4 | + | ||
| 5 | +| Product | Supported| | ||
| 6 | +| ------------------------------------------------------------ | :------: | | ||
| 7 | +|<term>Atlas A3 training products</term> | √ | | ||
| 8 | +|<term>Atlas A2 training products</term> | √ | | ||
| 9 | + | ||
| 10 | +## Function | ||
| 11 | + | ||
| 12 | +Allocates a special tensor whose device configuration is NPU but whose actual memory resides on the host. | ||
| 13 | + | ||
| 14 | +## Definition File | ||
| 15 | + | ||
| 16 | +torch_npu\csrc\core\npu\NPUFormat.h | ||
| 17 | + | ||
| 18 | +## Prototype | ||
| 19 | + | ||
| 20 | +```cpp | ||
| 21 | +at::Tensor empty_with_swapped_memory(c10::IntArrayRef size, c10::optional<at::ScalarType> dtype_opt, c10::optional<c10::Device> device_opt) | ||
| 22 | +``` | ||
| 23 | + | ||
| 24 | +## Parameters | ||
| 25 | + | ||
| 26 | +- **`size`** (`c10::IntArrayRef`): Required. Shape of the tensor to be created. | ||
| 27 | +- **`dtype_opt`** (`c10::optional<at::ScalarType>`): Optional. Data type of the tensor to be created. If the value is `c10::nullopt`, the global default data type is used. | ||
| 28 | +- **`device_opt`** (`c10::optional<c10::Device>`): Optional. Device configuration of the tensor to be created. If the value is `c10::nullopt`, the current default device is used. | ||
| 29 | + | ||
| 30 | +## Return Values | ||
| 31 | + | ||
| 32 | +`at::Tensor` | ||
| 33 | + | ||
| 34 | +Special `Tensor` generated. | ||
| 35 | + | ||
| 36 | +## Constraints | ||
| 37 | + | ||
| 38 | +- This API does not support graph mode. | ||
| 39 | + | ||
| 40 | +- Currently, the special `Tensor` allocated by this API supports only the following operators:<br> | ||
| 41 | +`aten::fill_`<br> | ||
| 42 | +`aten::zero_`<br> | ||
| 43 | +`aten::mul_`<br> | ||
| 44 | +`npu_apply_adam_w`<br> | ||
| 45 | +`npu_hans_encode`<br> | ||
| 46 | +`npu_hans_decode`<br> | ||
| 47 | + | ||
| 48 | +- When the installed CANN version is 8.5.0 or later and the Ascend HDK version is 26.0.rc1 or later, the special tensor allocated by this API supports direct printing. | ||
| 49 | +- When the installed CANN version is earlier than 8.5.0 or the Ascend HDK version is earlier than 26.0.rc1, the special tensor allocated by this API does not support direct printing. In this case, a warning log is printed. To view the value, convert it to a regular tensor through `mul_` before printing. | ||
| @@ -0,0 +1,34 @@ | |||
| 1 | +# c10_npu::NPUStreamGuard::NPUStreamGuard | ||
| 2 | + | ||
| 3 | +## Supported Products | ||
| 4 | + | ||
| 5 | +| Product | Supported| | ||
| 6 | +| ------------------------------------------------------------ | :------: | | ||
| 7 | +|<term>Atlas A3 training products</term> | √ | | ||
| 8 | +|<term>Atlas A2 training products</term> | √ | | ||
| 9 | + | ||
| 10 | +## Function | ||
| 11 | + | ||
| 12 | +Creates a stream guard. This sets the current NPU device to the device associated with the provided stream, and configures the current NPU stream. Upon destruction, the guard automatically restores the NPU device and stream to their pre-construction states. This ensures the original context is recovered when the scope exits. | ||
| 13 | + | ||
| 14 | +## Definition File | ||
| 15 | + | ||
| 16 | +torch_npu\csrc\core\npu\NPUGuard.h | ||
| 17 | + | ||
| 18 | +## Prototype | ||
| 19 | + | ||
| 20 | +```cpp | ||
| 21 | +c10_npu::NPUStreamGuard::NPUStreamGuard(c10::Stream stream) | ||
| 22 | +``` | ||
| 23 | + | ||
| 24 | +## Parameters | ||
| 25 | + | ||
| 26 | +`stream` (`c10::Stream`): Required. Stream managed by the guard. | ||
| 27 | + | ||
| 28 | +## Return Values | ||
| 29 | + | ||
| 30 | +None | ||
| 31 | + | ||
| 32 | +## Constraints | ||
| 33 | + | ||
| 34 | +None | ||
| @@ -0,0 +1,36 @@ | |||
| 1 | +# c10_npu::NPUStreamGuard::current_device | ||
| 2 | + | ||
| 3 | +## Supported Products | ||
| 4 | + | ||
| 5 | +| Product | Supported| | ||
| 6 | +| ------------------------------------------------------------ | :------: | | ||
| 7 | +|<term>Atlas A3 training products</term> | √ | | ||
| 8 | +|<term>Atlas A2 training products</term> | √ | | ||
| 9 | + | ||
| 10 | +## Function | ||
| 11 | + | ||
| 12 | +Returns the current NPU device set by this stream guard. | ||
| 13 | + | ||
| 14 | +## Definition File | ||
| 15 | + | ||
| 16 | +torch_npu\csrc\core\npu\NPUGuard.h | ||
| 17 | + | ||
| 18 | +## Prototype | ||
| 19 | + | ||
| 20 | +```cpp | ||
| 21 | +c10::Device c10_npu::NPUStreamGuard::current_device() const | ||
| 22 | +``` | ||
| 23 | + | ||
| 24 | +## Parameters | ||
| 25 | + | ||
| 26 | +None | ||
| 27 | + | ||
| 28 | +## Return Values | ||
| 29 | + | ||
| 30 | +`c10::Device` | ||
| 31 | + | ||
| 32 | +The current device. | ||
| 33 | + | ||
| 34 | +## Constraints | ||
| 35 | + | ||
| 36 | +None | ||
| @@ -0,0 +1,36 @@ | |||
| 1 | +# c10_npu::NPUStreamGuard::current_stream | ||
| 2 | + | ||
| 3 | +## Supported Products | ||
| 4 | + | ||
| 5 | +| Product | Supported| | ||
| 6 | +| ------------------------------------------------------------ | :------: | | ||
| 7 | +|<term>Atlas A3 training products</term> | √ | | ||
| 8 | +|<term>Atlas A2 training products</term> | √ | | ||
| 9 | + | ||
| 10 | +## Function | ||
| 11 | + | ||
| 12 | +Returns the latest NPU stream set by this stream guard. | ||
| 13 | + | ||
| 14 | +## Definition File | ||
| 15 | + | ||
| 16 | +torch_npu\csrc\core\npu\NPUGuard.h | ||
| 17 | + | ||
| 18 | +## Prototype | ||
| 19 | + | ||
| 20 | +```cpp | ||
| 21 | +c10_npu::NPUStream c10_npu::NPUStreamGuard::current_stream() const | ||
| 22 | +``` | ||
| 23 | + | ||
| 24 | +## Parameters | ||
| 25 | + | ||
| 26 | +None | ||
| 27 | + | ||
| 28 | +## Return Values | ||
| 29 | + | ||
| 30 | +`c10_npu::NPUStream` | ||
| 31 | + | ||
| 32 | +The currently managed stream. | ||
| 33 | + | ||
| 34 | +## Constraints | ||
| 35 | + | ||
| 36 | +None | ||
| @@ -0,0 +1,36 @@ | |||
| 1 | +# c10_npu::NPUStreamGuard::original_device | ||
| 2 | + | ||
| 3 | +## Supported Products | ||
| 4 | + | ||
| 5 | +| Product | Supported| | ||
| 6 | +| ------------------------------------------------------------ | :------: | | ||
| 7 | +|<term>Atlas A3 training products</term> | √ | | ||
| 8 | +|<term>Atlas A2 training products</term> | √ | | ||
| 9 | + | ||
| 10 | +## Function | ||
| 11 | + | ||
| 12 | +Returns the NPU device captured during guard construction. | ||
| 13 | + | ||
| 14 | +## Definition File | ||
| 15 | + | ||
| 16 | +torch_npu\csrc\core\npu\NPUGuard.h | ||
| 17 | + | ||
| 18 | +## Prototype | ||
| 19 | + | ||
| 20 | +```cpp | ||
| 21 | +c10::Device c10_npu::NPUStreamGuard::original_device() const | ||
| 22 | +``` | ||
| 23 | + | ||
| 24 | +## Parameters | ||
| 25 | + | ||
| 26 | +None | ||
| 27 | + | ||
| 28 | +## Return Values | ||
| 29 | + | ||
| 30 | +`c10::Device` | ||
| 31 | + | ||
| 32 | +NPU device captured during guard construction. | ||
| 33 | + | ||
| 34 | +## Constraints | ||
| 35 | + | ||
| 36 | +None | ||
| @@ -0,0 +1,36 @@ | |||
| 1 | +# c10_npu::NPUStreamGuard::original_stream | ||
| 2 | + | ||
| 3 | +## Supported Products | ||
| 4 | + | ||
| 5 | +| Product | Supported| | ||
| 6 | +| ------------------------------------------------------------ | :------: | | ||
| 7 | +|<term>Atlas A3 training products</term> | √ | | ||
| 8 | +|<term>Atlas A2 training products</term> | √ | | ||
| 9 | + | ||
| 10 | +## Function | ||
| 11 | + | ||
| 12 | +Returns the NPU stream set during guard construction. | ||
| 13 | + | ||
| 14 | +## Definition File | ||
| 15 | + | ||
| 16 | +torch_npu\csrc\core\npu\NPUGuard.h | ||
| 17 | + | ||
| 18 | +## Prototype | ||
| 19 | + | ||
| 20 | +```cpp | ||
| 21 | +c10_npu::NPUStream c10_npu::NPUStreamGuard::original_stream() const | ||
| 22 | +``` | ||
| 23 | + | ||
| 24 | +## Parameters | ||
| 25 | + | ||
| 26 | +None | ||
| 27 | + | ||
| 28 | +## Return Values | ||
| 29 | + | ||
| 30 | +`c10_npu::NPUStream` | ||
| 31 | + | ||
| 32 | +Stream set during guard construction. | ||
| 33 | + | ||
| 34 | +## Constraints | ||
| 35 | + | ||
| 36 | +None | ||
| @@ -0,0 +1,34 @@ | |||
| 1 | +# c10_npu::NPUStreamGuard::reset_stream | ||
| 2 | + | ||
| 3 | +## Supported Products | ||
| 4 | + | ||
| 5 | +| Product | Supported| | ||
| 6 | +| ------------------------------------------------------------ | :------: | | ||
| 7 | +|<term>Atlas A3 training products</term> | √ | | ||
| 8 | +|<term>Atlas A2 training products</term> | √ | | ||
| 9 | + | ||
| 10 | +## Function | ||
| 11 | + | ||
| 12 | +Sets a new NPU stream for the guard. This function first restores the current stream and device to their original states. It then sets the current device to the device associated with the provided stream and sets the current NPU stream on that device to the provided stream. | ||
| 13 | + | ||
| 14 | +## Definition File | ||
| 15 | + | ||
| 16 | +torch_npu\csrc\core\npu\NPUGuard.h | ||
| 17 | + | ||
| 18 | +## Prototype | ||
| 19 | + | ||
| 20 | +```cpp | ||
| 21 | +void c10_npu::NPUStreamGuard::reset_stream(c10::Stream stream) | ||
| 22 | +``` | ||
| 23 | + | ||
| 24 | +## Parameters | ||
| 25 | + | ||
| 26 | +**`stream`** (`c10::Stream`): Required. Stream to be managed by the guard. | ||
| 27 | + | ||
| 28 | +## Return Values | ||
| 29 | + | ||
| 30 | +None | ||
| 31 | + | ||
| 32 | +## Constraints | ||
| 33 | + | ||
| 34 | +`stream` must be an NPU stream (a `c10::Stream` created by an NPU device). Otherwise, the behavior is undefined. | ||
| @@ -0,0 +1,22 @@ | |||
| 1 | +# c10_npu::NPUStreamGuard | ||
| 2 | + | ||
| 3 | +## Supported Products | ||
| 4 | + | ||
| 5 | +| Product | Supported| | ||
| 6 | +| ------------------------------------------------------------ | :------: | | ||
| 7 | +|<term>Atlas A3 training products</term> | √ | | ||
| 8 | +|<term>Atlas A2 training products</term> | √ | | ||
| 9 | + | ||
| 10 | +## Function | ||
| 11 | + | ||
| 12 | +Manages the NPU device stream within its execution scope. This guard is identical to `c10::cuda::CUDAStreamGuard`. | ||
| 13 | + | ||
| 14 | +## Definition File | ||
| 15 | + | ||
| 16 | +torch_npu\csrc\core\npu\NPUGuard.h | ||
| 17 | + | ||
| 18 | +## Prototype | ||
| 19 | + | ||
| 20 | +```cpp | ||
| 21 | +struct c10_npu::NPUStreamGuard | ||
| 22 | +``` | ||
| @@ -0,0 +1,34 @@ | |||
| 1 | +# c10_npu::stream_synchronize | ||
| 2 | + | ||
| 3 | +## Supported Products | ||
| 4 | + | ||
| 5 | +| Product | Supported| | ||
| 6 | +| ------------------------------------------------------------ | :------: | | ||
| 7 | +|<term>Atlas A3 training products</term> | √ | | ||
| 8 | +|<term>Atlas A2 training products</term> | √ | | ||
| 9 | + | ||
| 10 | +## Function | ||
| 11 | + | ||
| 12 | +Synchronizes an NPU stream. This function is identical to `c10::cuda::stream_synchronize`. | ||
| 13 | + | ||
| 14 | +## Definition File | ||
| 15 | + | ||
| 16 | +torch_npu\csrc\core\npu\NPUFunctions.h | ||
| 17 | + | ||
| 18 | +## Prototype | ||
| 19 | + | ||
| 20 | +```cpp | ||
| 21 | +void stream_synchronize(aclrtStream stream) | ||
| 22 | +``` | ||
| 23 | + | ||
| 24 | +## Parameters | ||
| 25 | + | ||
| 26 | +**`stream`** (`aclrtStream`): Required. Stream to be synchronized. | ||
| 27 | + | ||
| 28 | +## Return Values | ||
| 29 | + | ||
| 30 | +None | ||
| 31 | + | ||
| 32 | +## Constraints | ||
| 33 | + | ||
| 34 | +None | ||
| @@ -0,0 +1,38 @@ | |||
| 1 | +# c10d_npu::ProcessGroupHCCL::batch_isend_irecv | ||
| 2 | + | ||
| 3 | +## Supported Products | ||
| 4 | + | ||
| 5 | +| Product | Supported| | ||
| 6 | +| ------------------------------------------------------------ | :------: | | ||
| 7 | +|<term>Atlas A3 training products</term> | √ | | ||
| 8 | +|<term>Atlas A2 training products</term> | √ | | ||
| 9 | + | ||
| 10 | +## Function | ||
| 11 | + | ||
| 12 | +Sends or receives a batch of tensors. Asynchronously processes each point-to-point (P2P) operation in the sequence. Returns the corresponding request handle. | ||
| 13 | + | ||
| 14 | +## Definition File | ||
| 15 | + | ||
| 16 | +torch_npu\csrc\distributed\ProcessGroupHCCL.hpp | ||
| 17 | + | ||
| 18 | +## Prototype | ||
| 19 | + | ||
| 20 | +```cpp | ||
| 21 | +c10::intrusive_ptr<c10d::Work> batch_isend_irecv(std::vector<std::string>& op_type, std::vector<at::Tensor>& tensors, std::vector<uint32_t>& remote_rank_list) | ||
| 22 | +``` | ||
| 23 | + | ||
| 24 | +## Parameters | ||
| 25 | + | ||
| 26 | +- **`op_type`** (`std::vector<std::string>&`): Required. Operation sequence. Each element in the vector is either `isend` or `irecv`, specifying the communication operation type for the corresponding tensor. | ||
| 27 | +- **`tensors`** (`std::vector<at::Tensor>&`): Required. Tensors to send or receive. The number of tensors must match the number of operations in `op_type`. | ||
| 28 | +- **`remote_rank_list`** (`std::vector<uint32_t>`): Required. Global rank IDs of the peer processes. The number of rank IDs must match the number of operations in `op_type`. | ||
| 29 | + | ||
| 30 | +## Return Values | ||
| 31 | + | ||
| 32 | +`c10::intrusive_ptr<c10d::Work>` | ||
| 33 | + | ||
| 34 | +Asynchronous `isend` and `irecv` operations. Because these operations execute asynchronously, a work handle is returned. Before reading any tensor data, call `work.wait()` to ensure that `batch_isend_irecv` is complete. | ||
| 35 | + | ||
| 36 | +## Constraints | ||
| 37 | + | ||
| 38 | +None | ||
| @@ -0,0 +1,44 @@ | |||
| 1 | +# c10d_npu::ProcessGroupHCCL | ||
| 2 | + | ||
| 3 | +## Supported Products | ||
| 4 | + | ||
| 5 | +| Product | Supported| | ||
| 6 | +| ------------------------------------------------------------ | :------: | | ||
| 7 | +|<term>Atlas A3 training products</term> | √ | | ||
| 8 | +|<term>Atlas A2 training products</term> | √ | | ||
| 9 | + | ||
| 10 | +## Function | ||
| 11 | + | ||
| 12 | +Inherits from `c10d::Backend` and implements the `HCCL` backend APIs for communication operators. | ||
| 13 | + | ||
| 14 | +## Definition File | ||
| 15 | + | ||
| 16 | +torch_npu\csrc\distributed\ProcessGroupHCCL.hpp | ||
| 17 | + | ||
| 18 | +## Prototype | ||
| 19 | + | ||
| 20 | +```cpp | ||
| 21 | +class c10d_npu::ProcessGroupHCCL | ||
| 22 | +``` | ||
| 23 | + | ||
| 24 | +## Constraints | ||
| 25 | + | ||
| 26 | +- To achieve optimal concurrency and performance, all HCCL communication functions provided by this class are asynchronous. You must call `WorkHCCL::wait()` or `WorkHCCL::synchronize()` to ensure that task execution is complete. | ||
| 27 | +- This class inherits from the native `c10d::Backend`. For details about the basic communication operator APIs, see the [PyTorch Distributed Documentation](https://docs.pytorch.org/docs/stable/distributed.html). They are not repeated in this document. The supported basic communication operators are as follows:<br> | ||
| 28 | +broadcast<br> | ||
| 29 | +allreduce<br> | ||
| 30 | +allreduce_coalesced<br> | ||
| 31 | +reduce<br> | ||
| 32 | +allgather<br> | ||
| 33 | +allgather_togather<br> | ||
| 34 | +allgather_into_tensor_coalesced<br> | ||
| 35 | +reduce_scatter<br> | ||
| 36 | +reduce_scatter_tensor_coalesced<br> | ||
| 37 | +barrier<br> | ||
| 38 | +gather<br> | ||
| 39 | +scatter<br> | ||
| 40 | +send<br> | ||
| 41 | +recv<br> | ||
| 42 | +recv_anysource<br> | ||
| 43 | +alltoall_base<br> | ||
| 44 | +alltoall<br> | ||
| @@ -0,0 +1,67 @@ | |||
| 1 | +# Supported Deterministic Computing APIs | ||
| 2 | + | ||
| 3 | +## Overview | ||
| 4 | + | ||
| 5 | +When training with the PyTorch framework, some operators may produce non-deterministic results during computation. If deterministic output is required, enable deterministic computation. When deterministic computation is enabled, performing the same operation with the same input on the same hardware and software produces the same output every time. | ||
| 6 | + | ||
| 7 | +> [!NOTE] | ||
| 8 | +> | ||
| 9 | +> - The configuration method for deterministic computation must run in the same main process as the target network or operator to be fixed. In some model scripts, `main()` and the training network do not run in the same process. | ||
| 10 | +> - Currently, the deterministic state can be configured only once within the same thread. If it is configured multiple times, only the first effective configuration takes effect, and subsequent configurations do not take effect.<br> | ||
| 11 | +> Effective configuration: After the deterministic state is set, at least one operator task must actually be dispatched and executed. If the deterministic state is set without dispatching any operator, only the deterministic variable is enabled and the setting is not applied to any operator. This is because the framework cannot determine which operator requires deterministic computation until an operator is executed.<br> | ||
| 12 | +> Solution: | ||
| 13 | +> 1. Repeatedly configuring deterministic computation within a single thread is not recommended. | ||
| 14 | +> 2. This issue exists regardless of whether binary mode is enabled or disabled, and it will be resolved in a future release. | ||
| 15 | + | ||
| 16 | +## Usage | ||
| 17 | + | ||
| 18 | +For details about the usage and effects of deterministic computation, see the official documentation for [torch.use_deterministic_algorithms](https://pytorch.org/docs/stable/generated/torch.use_deterministic_algorithms.html#torch.use_deterministic_algorithms). This section describes only how to enable deterministic computation. | ||
| 19 | + | ||
| 20 | +> [!CAUTION] | ||
| 21 | +> Enabling the deterministic computation switch may degrade performance. | ||
| 22 | + | ||
| 23 | +1. Enable deterministic computation: | ||
| 24 | + | ||
| 25 | + ```python | ||
| 26 | + torch.use_deterministic_algorithms(True) | ||
| 27 | + ``` | ||
| 28 | + | ||
| 29 | +2. Verify whether the configuration is successful. | ||
| 30 | + | ||
| 31 | + 1. Run the following command to query the API configuration: | ||
| 32 | + | ||
| 33 | + ```python | ||
| 34 | + torch.are_deterministic_algorithms_enabled() | ||
| 35 | + ``` | ||
| 36 | + | ||
| 37 | + 2. The following output is displayed: | ||
| 38 | + | ||
| 39 | + ```python | ||
| 40 | + print(torch.are_deterministic_algorithms_enabled()) | ||
| 41 | + ``` | ||
| 42 | + | ||
| 43 | + During training, a return value of `True` from this API indicates that deterministic computation is enabled, whereas `False` indicates that it is disabled. | ||
| 44 | + | ||
| 45 | +## API Lists | ||
| 46 | + | ||
| 47 | +When using <term>Atlas A2 training products/Atlas A2 inference products</term> or <term>Atlas A3 training products/Atlas A3 inference products</term>, the APIs listed in [Table 1](#api-list-for-enabling-deterministic-computation-1) may produce non-deterministic results during computation. Enabling deterministic computation ensures that the computation results are deterministic. | ||
| 48 | + | ||
| 49 | +**Table 1** API list<a id="api-list-for-enabling-deterministic-computation-1"></a> | ||
| 50 | + | ||
| 51 | +| API | | ||
| 52 | +|-----| | ||
| 53 | +| `torch_npu.npu_convolution_transpose` | | ||
| 54 | +| `torch_npu.npu_linear` | | ||
| 55 | +| `torch_npu.npu_deformable_conv2d` | | ||
| 56 | + | ||
| 57 | +When using <term>Ascend 950DT</term>, the APIs listed in [Table 2](#api-list-for-enabling-deterministic-computation-2) may produce non-deterministic results during computation. Enabling deterministic computation ensures that the computation results are deterministic. | ||
| 58 | + | ||
| 59 | +**Table 2** API list<a id="api-list-for-enabling-deterministic-computation-2"></a> | ||
| 60 | + | ||
| 61 | +| API | | ||
| 62 | +|-----| | ||
| 63 | +| `torch_npu.npu_scatter_nd_update` | | ||
| 64 | +| `torch_npu.npu_scatter_nd_update_` | | ||
| 65 | +| `torch_npu.scatter_update` | | ||
| 66 | +| `torch_npu.scatter_update_` | | ||
| 67 | +| `torch_npu.npu_fusion_attention_grad` | | ||
| @@ -0,0 +1,31 @@ | |||
| 1 | +# (beta) torch.distributed.ProcessGroupHCCL | ||
| 2 | + | ||
| 3 | +## Supported Products | ||
| 4 | + | ||
| 5 | +| Product | Supported| | ||
| 6 | +| ------------------------------------------------------------ | :------: | | ||
| 7 | +|<term>Atlas A3 training products</term> | √ | | ||
| 8 | +|<term>Atlas A2 training products</term> | √ | | ||
| 9 | +|<term>Atlas inference products</term> | √ | | ||
| 10 | +|<term>Atlas training products</term> | √ | | ||
| 11 | + | ||
| 12 | +## Function | ||
| 13 | + | ||
| 14 | +Creates and returns a `ProcessGroupHCCL` object. | ||
| 15 | + | ||
| 16 | +## Prototype | ||
| 17 | + | ||
| 18 | +```python | ||
| 19 | +torch.distributed.ProcessGroupHCCL(store, rank, size, timeout) -> ProcessGroup | ||
| 20 | +``` | ||
| 21 | + | ||
| 22 | +## Parameters | ||
| 23 | + | ||
| 24 | +- **`store`** (`torch.distributed.distributed_c10d.PrefixStore`): `PrefixStore` object created using its constructor. | ||
| 25 | +- **`rank`**: Rank ID of the current node. | ||
| 26 | +- **`size`**: Total number of communication nodes. | ||
| 27 | +- **`timeout`**: Communication timeout used to detect node disconnection. The default value is `1800` (seconds). | ||
| 28 | + | ||
| 29 | +## Return Values | ||
| 30 | + | ||
| 31 | +`ProcessGroup` | ||
| @@ -0,0 +1,35 @@ | |||
| 1 | +# (beta) torch.distributed.is_hccl_available | ||
| 2 | + | ||
| 3 | +## Supported Products | ||
| 4 | + | ||
| 5 | +| Product | Supported| | ||
| 6 | +| ------------------------------------------------------------ | :------: | | ||
| 7 | +|<term>Atlas A3 training products</term> | √ | | ||
| 8 | +|<term>Atlas A2 training products</term> | √ | | ||
| 9 | +|<term>Atlas inference products</term> | √ | | ||
| 10 | +|<term>Atlas training products</term> | √ | | ||
| 11 | + | ||
| 12 | +## Function | ||
| 13 | + | ||
| 14 | +Determines whether the `HCCL` communication backend is available, similar to `torch.distributed.is_nccl_available`. For details, see [https://pytorch.org/docs/stable/distributed.html\#torch.distributed.is_nccl_available](https://pytorch.org/docs/stable/distributed.html#torch.distributed.is_nccl_available). | ||
| 15 | + | ||
| 16 | +## Prototype | ||
| 17 | + | ||
| 18 | +```python | ||
| 19 | +torch.distributed.is_hccl_available() | ||
| 20 | +``` | ||
| 21 | + | ||
| 22 | +## Return Values | ||
| 23 | + | ||
| 24 | +A `bool` value: `True` (available) or `False` (unavailable). | ||
| 25 | + | ||
| 26 | +## Example | ||
| 27 | + | ||
| 28 | +```python | ||
| 29 | +import torch | ||
| 30 | +import torch_npu | ||
| 31 | + | ||
| 32 | +print(torch.distributed.is_hccl_available()) | ||
| 33 | + | ||
| 34 | +True | ||
| 35 | +``` | ||
| @@ -0,0 +1,70 @@ | |||
| 1 | +# (beta) torch_npu.distributed.all_gather_into_tensor_uneven | ||
| 2 | + | ||
| 3 | +## Supported Products | ||
| 4 | + | ||
| 5 | +| Product | Supported| | ||
| 6 | +| ------------------------------------------------------------ | :------: | | ||
| 7 | +|<term>Atlas A2 training products/Atlas 800I A2 inference products</term> | √ | | ||
| 8 | +|<term>Atlas inference products</term> | √ | | ||
| 9 | + | ||
| 10 | +## Function | ||
| 11 | + | ||
| 12 | +Extends the native [torch.distributed.all_gather_into_tensor](https://pytorch.org/docs/stable/distributed.html#torch.distributed.all_gather_into_tensor) API by supporting zero-copy and uneven tensor splitting in `torch_npu.distributed.all_gather_into_tensor_uneven`. | ||
| 13 | + | ||
| 14 | +## Prototype | ||
| 15 | + | ||
| 16 | +```python | ||
| 17 | +torch_npu.distributed.all_gather_into_tensor_uneven(output, input, output_split_sizes =None, group=None, async_op=False) -> torch.distributed.distributed_c10d.Work | ||
| 18 | +``` | ||
| 19 | + | ||
| 20 | +## Parameters | ||
| 21 | + | ||
| 22 | +- **`output`** (`Tensor`): Output tensor used to receive the computation result. | ||
| 23 | +- **`input`** (`Tensor`): Input tensor used to provide computation data. No special constraints are imposed on the shape of `input`. | ||
| 24 | +- **`output_split_sizes`** (`List[int]`): Split sizes along dimension 0 of the `output` tensor. The default value is `None`. The number of elements must match the size of the `group`, and the sum must equal the size of dimension 0 of `output`. | ||
| 25 | + - If the sum of `output_split_sizes` does not match the size of dimension 0 of `output`, a `RuntimeError` is raised: "Split sizes doesn't match total dim 0 size." | ||
| 26 | + - If the number of elements in `output_split_sizes` does not match the size of `group`, a `RuntimeError` is raised: "Number of tensor splits not equal to group size." | ||
| 27 | +- **`group`** (`torch.distributed.distributed_c10d.ProcessGroup`): The process group for distributed communication. The default value is `None`. | ||
| 28 | +- **`async_op`** (`bool`): Specifies whether to execute the operation asynchronously. The default value is `False`. | ||
| 29 | + | ||
| 30 | +## Return Values | ||
| 31 | + | ||
| 32 | +The shape of `output` is the shape obtained by concatenating the `input` tensors across all ranks. | ||
| 33 | + | ||
| 34 | +## Constraints | ||
| 35 | + | ||
| 36 | +- This API can be used only in single-server scenarios. | ||
| 37 | + | ||
| 38 | +- The sum of `output_split_sizes` must equal the size of dimension 0 of `output`. The number of elements in `output_split_sizes` must match the size of the `group`. | ||
| 39 | + | ||
| 40 | +## Example | ||
| 41 | + | ||
| 42 | +Create the `test.py` file and save it. | ||
| 43 | + | ||
| 44 | +```python | ||
| 45 | +import os | ||
| 46 | +import torch | ||
| 47 | +import torch_npu | ||
| 48 | +import torch.distributed as dist | ||
| 49 | + | ||
| 50 | +dist.init_process_group(backend="hccl") | ||
| 51 | +rank = int(os.getenv('LOCAL_RANK')) | ||
| 52 | +torch.npu.set_device(rank) | ||
| 53 | + | ||
| 54 | +output_split_sizes = [2, 3] | ||
| 55 | +input_tensors = [torch.tensor([1, 2], dtype=torch.int32).npu(), torch.tensor([4, 5, 6], dtype=torch.int32).npu()] | ||
| 56 | +output_tensor = torch.zeros(sum(output_split_sizes), dtype=torch.int32).npu() | ||
| 57 | + | ||
| 58 | +torch_npu.distributed.all_gather_into_tensor_uneven( | ||
| 59 | + output_tensor, | ||
| 60 | + input_tensors[rank], | ||
| 61 | + output_split_sizes=output_split_sizes, | ||
| 62 | + async_op=False | ||
| 63 | +) | ||
| 64 | +``` | ||
| 65 | + | ||
| 66 | +Run the following command: | ||
| 67 | + | ||
| 68 | +```bash | ||
| 69 | +torchrun --nproc-per-node=2 test.py | ||
| 70 | +``` | ||
| @@ -0,0 +1,63 @@ | |||
| 1 | +# (beta) torch_npu.distributed.reinit_process_group | ||
| 2 | + | ||
| 3 | +## Supported Products | ||
| 4 | + | ||
| 5 | +| Product | Supported| | ||
| 6 | +| ------------------------------------------------------------ | :------: | | ||
| 7 | +|<term>Atlas A3 training products</term> | √ | | ||
| 8 | +|<term>Atlas A2 training products</term> | √ | | ||
| 9 | + | ||
| 10 | +## Function | ||
| 11 | + | ||
| 12 | +Rebuilds the `ProcessGroup` for collective communication. | ||
| 13 | + | ||
| 14 | +## Prototype | ||
| 15 | + | ||
| 16 | +```python | ||
| 17 | +torch_npu.distributed.reinit_process_group(group: Optional[ProcessGroup] = None, rebuild_link: bool = True) -> None | ||
| 18 | +``` | ||
| 19 | + | ||
| 20 | +## Parameters | ||
| 21 | + | ||
| 22 | +- **`group`** (`Optional[ProcessGroup]`): Optional. Communication group. The default value is `None`. | ||
| 23 | +- **`rebuild_link`** (`bool`): Optional. When set to `True`, the framework destroys and rebuilds the communication links created by the current `ProcessGroupHCCL` instance. When set to `False`, execution reuses the existing communication links. The default value is `True`. | ||
| 24 | + | ||
| 25 | +## Return Values | ||
| 26 | + | ||
| 27 | +None | ||
| 28 | + | ||
| 29 | +## Constraints | ||
| 30 | + | ||
| 31 | +The specified input device must be valid. | ||
| 32 | + | ||
| 33 | +## Example | ||
| 34 | + | ||
| 35 | +```python | ||
| 36 | +import os | ||
| 37 | +import torch | ||
| 38 | +import torch.distributed as dist | ||
| 39 | +import multiprocessing as mp | ||
| 40 | +import torch_npu | ||
| 41 | +def _do_allreduce(rank, world_size): | ||
| 42 | + os.environ['MASTER_ADDR'] = '127.0.0.1' | ||
| 43 | + os.environ['MASTER_PORT'] = '29688' | ||
| 44 | + torch.npu.set_device(rank) | ||
| 45 | + dist.init_process_group(backend='hccl', | ||
| 46 | + world_size=world_size, | ||
| 47 | + rank=rank) | ||
| 48 | + # Rebuild a group | ||
| 49 | + torch_npu.distributed.reinit_process_group() | ||
| 50 | + a = torch.ones(2,2,device=f"npu:{rank}") | ||
| 51 | + dist.all_reduce(a) | ||
| 52 | +def _multiprocess(world_size,f): | ||
| 53 | + ctx = mp.get_context('spawn') | ||
| 54 | + ps = [] | ||
| 55 | + for i in range(world_size): | ||
| 56 | + p = ctx.Process(target=f, args=(i,world_size)) | ||
| 57 | + p.start() | ||
| 58 | + ps.append(p) | ||
| 59 | + for p in ps: | ||
| 60 | + p.join() | ||
| 61 | +if __name__ == '__main__': | ||
| 62 | + _multiprocess(4, _do_allreduce) | ||
| 63 | +``` | ||
| @@ -0,0 +1 @@ | |||
| 1 | +# Distributed | ||
| @@ -0,0 +1,45 @@ | |||
| 1 | +# Distributed APIs | ||
| 2 | + | ||
| 3 | +This section describes the adapted distributed APIs that provide parallel computing capabilities. | ||
| 4 | + | ||
| 5 | +**Table 1** Distributed APIs | ||
| 6 | + | ||
| 7 | +<a name="table2069619331171"></a> | ||
| 8 | +<table><thead align="left"><tr id="row86962336177"><th class="cellrowborder" valign="top" width="50%" id="mcps1.2.3.1.1"><p id="p2696143312179"><a name="p2696143312179"></a><a name="p2696143312179"></a>API</p> | ||
| 9 | +</th> | ||
| 10 | +<th class="cellrowborder" valign="top" width="50%" id="mcps1.2.3.1.2"><p id="p669613318178"><a name="p669613318178"></a><a name="p669613318178"></a>Description</p> | ||
| 11 | +</th> | ||
| 12 | +</tr> | ||
| 13 | +</thead> | ||
| 14 | +<tbody><tr id="row669611337171"><td class="cellrowborder" valign="top" width="50%" headers="mcps1.2.3.1.1 "><p id="p46966336179"><a name="p46966336179"></a><a name="p46966336179"></a><a href="(beta)torch-distributed-is_hccl_available.md">(beta) torch.distributed.is_hccl_available</a></p> | ||
| 15 | +</td> | ||
| 16 | +<td class="cellrowborder" valign="top" width="50%" headers="mcps1.2.3.1.2 "><p id="p176961133151712"><a name="p176961133151712"></a><a name="p176961133151712"></a>Determines whether the <code>HCCL</code> communication backend is available, similar to <code>torch.distributed.is_nccl_available</code>.</p> | ||
| 17 | +</td> | ||
| 18 | +</tr> | ||
| 19 | +<tr id="row186961833171719"><td class="cellrowborder" valign="top" width="50%" headers="mcps1.2.3.1.1 "><p id="p1669719330179"><a name="p1669719330179"></a><a name="p1669719330179"></a><a href="torch-distributed-distributed_c10d.md">torch.distributed.distributed_c10d._world.default_pg._get_backend(torch.device("npu")).get_hccl_comm_name</a></p> | ||
| 20 | +</td> | ||
| 21 | +<td class="cellrowborder" valign="top" width="50%" headers="mcps1.2.3.1.2 "><p id="p5697183351712"><a name="p5697183351712"></a><a name="p5697183351712"></a>Obtains the name of the collective communication domain from the initialized domain.</p> | ||
| 22 | +</td> | ||
| 23 | +</tr> | ||
| 24 | +<tr id="row174513178290"><td class="cellrowborder" valign="top" width="50%" headers="mcps1.2.3.1.1 "><p id="p204514170294"><a name="p204514170294"></a><a name="p204514170294"></a><a href="(beta)torch-distributed-ProcessGroupHCCL.md">(beta) torch.distributed.ProcessGroupHCCL</a></p> | ||
| 25 | +</td> | ||
| 26 | +<td class="cellrowborder" valign="top" width="50%" headers="mcps1.2.3.1.2 "><p id="en-us_topic_0000001788457948_en-us_topic_0000001718962912_p173841155158"><a name="en-us_topic_0000001788457948_en-us_topic_0000001718962912_p173841155158"></a><a name="en-us_topic_0000001788457948_en-us_topic_0000001718962912_p173841155158"></a>Creates and returns a <code>ProcessGroupHCCL</code> object.</p> | ||
| 27 | +</td> | ||
| 28 | +</tr> | ||
| 29 | +<tr id="row193021022143117"><td class="cellrowborder" valign="top" width="50%" headers="mcps1.2.3.1.1 "><p id="p1530342217311"><a name="p1530342217311"></a><a name="p1530342217311"></a><a href="(beta)torch_npu-distributed-reinit_process_group.md">(beta) torch_npu.distributed.reinit_process_group</a></p> | ||
| 30 | +</td> | ||
| 31 | +<td class="cellrowborder" valign="top" width="50%" headers="mcps1.2.3.1.2 "><p id="p03031822133115"><a name="p03031822133115"></a><a name="p03031822133115"></a>Rebuilds the <code>ProcessGroup</code> for collective communication.</p> | ||
| 32 | +</td> | ||
| 33 | +</tr> | ||
| 34 | +<tr id="row6513134685111"><td class="cellrowborder" valign="top" width="50%" headers="mcps1.2.3.1.1 "><p id="p451334611515"><a name="p451334611515"></a><a name="p451334611515"></a><a href="torch_npu-distributed-reduce_scatter_tensor_uneven.md">(beta) torch_npu.distributed.reduce_scatter_tensor_uneven</a></p> | ||
| 35 | +</td> | ||
| 36 | +<td class="cellrowborder" valign="top" width="50%" headers="mcps1.2.3.1.2 "><p id="p651315465516"><a name="p651315465516"></a><a name="p651315465516"></a>Extends the native <code>torch.distributed.reduce_scatter_tensor</code> API by supporting zero-copy and uneven tensor splitting in <code>torch_npu.distributed.reduce_scatter_tensor_uneven</code>.</p> | ||
| 37 | +</td> | ||
| 38 | +</tr> | ||
| 39 | +<tr id="row125161049165116"><td class="cellrowborder" valign="top" width="50%" headers="mcps1.2.3.1.1 "><p id="p105171249165115"><a name="p105171249165115"></a><a name="p105171249165115"></a><a href="(beta)torch_npu-distributed-all_gather_into_tensor_uneven.md">(beta) torch_npu.distributed.all_gather_into_tensor_uneven</a></p> | ||
| 40 | +</td> | ||
| 41 | +<td class="cellrowborder" valign="top" width="50%" headers="mcps1.2.3.1.2 "><p id="p2517104919514"><a name="p2517104919514"></a><a name="p2517104919514"></a>Extends the native <code>torch.distributed.all_gather_into_tensor</code> API by supporting zero-copy and uneven tensor splitting in <code>torch_npu.distributed.all_gather_into_tensor_uneven</code>.</p> | ||
| 42 | +</td> | ||
| 43 | +</tr> | ||
| 44 | +</tbody> | ||
| 45 | +</table> | ||
| @@ -0,0 +1,78 @@ | |||
| 1 | +# torch.distributed.distributed_c10d._world.default_pg._get_backend(torch.device("npu")).get_hccl_comm_name | ||
| 2 | + | ||
| 3 | +## Supported Products | ||
| 4 | + | ||
| 5 | +| Product | Supported| | ||
| 6 | +| ------------------------------------------------------------ | :------: | | ||
| 7 | +|<term>Atlas A3 training products</term> | √ | | ||
| 8 | +|<term>Atlas A2 training products</term> | √ | | ||
| 9 | +|<term>Atlas inference products</term> | √ | | ||
| 10 | +|<term>Atlas training products</term> | √ | | ||
| 11 | + | ||
| 12 | +## Function | ||
| 13 | + | ||
| 14 | +Obtains the name of the collective communication domain from the initialized domain. | ||
| 15 | + | ||
| 16 | +## Prototype | ||
| 17 | + | ||
| 18 | +```python | ||
| 19 | +torch.distributed.distributed_c10d._world.default_pg._get_backend(torch.device("npu")).get_hccl_comm_name(rankid, init_comm=True) -> str | ||
| 20 | +``` | ||
| 21 | + | ||
| 22 | +Note: This method belongs to the PyTorch `ProcessGroup` class using the NPU backend. The `ProcessGroup` can be `default_pg` or a non-default group created by `torch.distributed.distributed_c10d.new_group`. | ||
| 23 | + | ||
| 24 | +> [!NOTICE] | ||
| 25 | +> When calling this API, ensure that the current device is correctly configured. | ||
| 26 | + | ||
| 27 | +## Parameters | ||
| 28 | + | ||
| 29 | +- **`rankid`** (`int`): Required. Rank ID of the device in the collective communication domain. The input is a global rank ID, which uniquely identifies a device across multiple servers. | ||
| 30 | + | ||
| 31 | +- **`init_comm`** (`bool`): Optional. Specifies whether to force initialization if `HCCL` is uninitialized. The default value is `True`. When set to `True`, calling `get_hccl_comm_name` initializes `HCCL` if it has not yet been initialized and returns the group name. When set to `False`, the API does not initialize `HCCL`. Instead, it returns an empty string if `HCCL` is not already initialized. | ||
| 32 | + | ||
| 33 | +>**Note**:<br> | ||
| 34 | +>`HCCL` initialization allocates memory resources, increasing memory usage. By default, 200 MB is allocated for the `send` buffer and 200 MB for the `receive` buffer, totaling 400 MB. The buffer size is controlled by the `HCCL_BUFFSIZE` environment variable. | ||
| 35 | + | ||
| 36 | +## Return Values | ||
| 37 | + | ||
| 38 | +`string` | ||
| 39 | + | ||
| 40 | +A `string` value representing the name of the collective communication domain. | ||
| 41 | + | ||
| 42 | +## Constraints | ||
| 43 | + | ||
| 44 | +- Ensure `init_process_group` has been called before using this API, and that the backend is set to `HCCL`. | ||
| 45 | +- The calling method differs between PyTorch version `2.1.0` (and later) and earlier versions. For details, see [Example](#section14459801435). | ||
| 46 | + | ||
| 47 | +## Example<a name="section14459801435"></a> | ||
| 48 | + | ||
| 49 | +```python | ||
| 50 | +import torch | ||
| 51 | +import torch_npu | ||
| 52 | +import torch.multiprocessing as mp | ||
| 53 | +import os | ||
| 54 | +from torch.distributed.distributed_c10d import _get_default_group | ||
| 55 | +import torch.distributed as dist | ||
| 56 | +def example(rank, world_size): | ||
| 57 | + torch.npu.set_device("npu:" + str(rank)) | ||
| 58 | + dist.init_process_group("hccl", rank=rank, world_size=world_size) | ||
| 59 | + default_pg = _get_default_group() | ||
| 60 | + if torch.__version__ > '2.0': | ||
| 61 | + hcomm_info = default_pg._get_backend(torch.device("npu")).get_hccl_comm_name(rank) | ||
| 62 | + else: | ||
| 63 | + hcomm_info = default_pg.get_hccl_comm_name(rank) | ||
| 64 | + print(hcomm_info) | ||
| 65 | +def main(): | ||
| 66 | + world_size = 2 | ||
| 67 | + mp.spawn(example, | ||
| 68 | + args=(world_size, ), | ||
| 69 | + nprocs=world_size, | ||
| 70 | + join=True) | ||
| 71 | +if __name__ == "__main__": | ||
| 72 | + os.environ["MASTER_ADDR"] = "localhost" | ||
| 73 | + os.environ["MASTER_PORT"] = "29505" | ||
| 74 | + main() | ||
| 75 | + | ||
| 76 | +group_name_0 | ||
| 77 | +group_name_0 | ||
| 78 | +``` | ||
| @@ -0,0 +1,69 @@ | |||
| 1 | +# (beta) torch_npu.distributed.reduce_scatter_tensor_uneven | ||
| 2 | + | ||
| 3 | +## Supported Products | ||
| 4 | + | ||
| 5 | +| Product | Supported| | ||
| 6 | +| ------------------------------------------------------------ | :------: | | ||
| 7 | +|<term>Atlas A2 training products/Atlas 800I A2 inference products</term> | √ | | ||
| 8 | +|<term>Atlas inference products</term> | √ | | ||
| 9 | + | ||
| 10 | +## Function | ||
| 11 | + | ||
| 12 | +Extends the native [torch.distributed.reduce_scatter_tensor](https://pytorch.org/docs/stable/distributed.html#torch.distributed.reduce_scatter_tensor) API by supporting zero-copy and uneven tensor splitting in <code>torch_npu.distributed.reduce_scatter_tensor_uneven</code>. | ||
| 13 | + | ||
| 14 | +## Prototype | ||
| 15 | + | ||
| 16 | +```python | ||
| 17 | +torch_npu.distributed.reduce_scatter_tensor_uneven(output, input, input_split_sizes =None, op=dist.ReduceOp.SUM, group=None, async_op=False) -> torch.distributed.distributed_c10d.Work | ||
| 18 | +``` | ||
| 19 | + | ||
| 20 | +## Parameters | ||
| 21 | + | ||
| 22 | +- **`output`** (`Tensor`): Required. Output tensor used to receive the computation result. | ||
| 23 | +- **`input`** (`Tensor`): Required. Input tensor providing the source data. The shape of `input` corresponds to the concatenation of `output` tensors across all ranks. | ||
| 24 | +- **`input_split_sizes`** (`List[int]`): Optional. Specifies the split sizes along dimension 0 of the `input` tensor. The default value is `None`. The number of elements must match the size of the `group`, and the sum of elements must equal the size of dimension 0 of `input`. | ||
| 25 | + - If the sum of `input_split_sizes` does not match the size of dimension 0 of `input`, a `RuntimeError` is raised: "Split sizes doesn't match total dim 0 size." | ||
| 26 | + - If the number of elements in `input_split_sizes` does not match the size of `group`, a `RuntimeError` is raised: "Number of tensor splits not equal to group size." | ||
| 27 | +- **`op`** (`torch._C._distributed_c10d.ReduceOp.ReduceOpType`): Optional. Reduction operator used to control the computation behavior. The default value is `dist.ReduceOp.SUM`. | ||
| 28 | +- **`group`** (`torch.distributed.distributed_c10d.ProcessGroup`): Optional. The process group for distributed communication. The default value is `None`. | ||
| 29 | +- **`async_op`** (`bool`): Optional. Specifies whether to execute the operation asynchronously. The default value is `False`. | ||
| 30 | + | ||
| 31 | +## Return Values | ||
| 32 | + | ||
| 33 | +A work handle used to track the asynchronous operation. The final result is written to `output`. | ||
| 34 | +No constraints are imposed on the shape of `output`. | ||
| 35 | + | ||
| 36 | +## Constraints | ||
| 37 | + | ||
| 38 | +- This API can be used only in single-server scenarios. | ||
| 39 | + | ||
| 40 | +- The sum of `input_split_sizes` must equal the size of dimension 0 of `input`. The number of elements in `input_split_sizes` must match the size of the `group`. | ||
| 41 | + | ||
| 42 | +## Example | ||
| 43 | + | ||
| 44 | +Create the test.py file and save it. | ||
| 45 | + | ||
| 46 | +```python | ||
| 47 | +import os | ||
| 48 | +import torch | ||
| 49 | +import torch_npu | ||
| 50 | +import torch.distributed as dist | ||
| 51 | +dist.init_process_group(backend="hccl") | ||
| 52 | +rank = int(os.getenv('LOCAL_RANK')) | ||
| 53 | +torch.npu.set_device(rank) | ||
| 54 | +input_split_sizes = [2, 3] | ||
| 55 | +input_tensor = torch.ones(sum(input_split_sizes), dtype=torch.int32).npu() | ||
| 56 | +output_tensor = torch.zeros(input_split_sizes[rank], dtype=torch.int32).npu() | ||
| 57 | +torch_npu.distributed.reduce_scatter_tensor_uneven( | ||
| 58 | + output_tensor, | ||
| 59 | + input_tensor, | ||
| 60 | + input_split_sizes=input_split_sizes, | ||
| 61 | + async_op=False | ||
| 62 | +) | ||
| 63 | +``` | ||
| 64 | + | ||
| 65 | +Run the following command: | ||
| 66 | + | ||
| 67 | +```bash | ||
| 68 | +torchrun --nproc-per-node=2 test.py | ||
| 69 | +``` | ||
| @@ -0,0 +1,16 @@ | |||
| 1 | +# Overview | ||
| 2 | + | ||
| 3 | +Provides function descriptions, prototypes, parameter descriptions, and call examples for custom TorchNPU APIs based on PyTorch versions such as 2.12.0, 2.11.0, 2.10.0, 2.9.0, and 2.7.1. | ||
| 4 | + | ||
| 5 | +The APIs provided by TorchNPU comply with the public API conventions defined by the PyTorch community. For details, see [Public API definition and documentation](https://github.com/pytorch/pytorch/wiki/Public-API-definition-and-documentation). The APIs described in this document are the public APIs of TorchNPU. Internal APIs may be modified or removed in future releases. Therefore, you are advised not to use them. If you must use them, submit an issue in the [Ascend community](https://gitcode.com/ascend/pytorch/issues) for assistance. | ||
| 6 | + | ||
| 7 | +TorchNPU integrates with PyTorch through monkey patching. Specifically, selected PyTorch APIs are dynamically replaced with TorchNPU implementations, allowing users to continue using familiar PyTorch APIs on Ascend NPUs. | ||
| 8 | + | ||
| 9 | +TorchNPU is developed using both C++ and Python. Currently, only Python APIs are officially exposed. C++ APIs are intended for internal use and are not recommended for users. | ||
| 10 | + | ||
| 11 | +Currently, some APIs are marked as beta APIs. Beta APIs are experimental and may exhibit unexpected behavior in certain scenarios. Exercise caution when using these APIs. We are committed to graduating beta APIs to stable APIs. However, before this process is complete, these APIs may still be changed as needed, including but not limited to parameter changes, renaming, and removal. | ||
| 12 | + | ||
| 13 | +By default, all custom APIs support all PyTorch versions compatible with the corresponding TorchNPU release. If an API does not support all compatible PyTorch versions, the restrictions are specified in the Constraints section of that API. | ||
| 14 | + | ||
| 15 | +> [!NOTE] | ||
| 16 | +> Some features of TorchNPU can be configured through environment variables. For details, see [Environment Variable Reference](https://gitcode.com/Ascend/pytorch/blob/v2.7.1-26.1.0/docs/en/environment_variable_reference/env_variable_list.md). | ||