[NPU] [DOC] fix typos, heading levels and terminology in NPU docs (#40402)

This commit is contained in:
amote-i
2026-09-20 15:15:58 +08:00
committed by GitHub
parent f4c256354c
commit 5c69e32abe
6 changed files with 17 additions and 16 deletions
@@ -37,7 +37,7 @@ sgl-kernel-npu/
│ ├── op_host/ # Host-side code (validation, tiling, launch) │ ├── op_host/ # Host-side code (validation, tiling, launch)
│ │ ├── <op_name>.cpp │ │ ├── <op_name>.cpp
│ │ └── tiling/ # Optional: tiling data │ │ └── tiling/ # Optional: tiling data
│ └── op_kernel/ # Device-side code (Ascend C kernel on AICore) │ └── op_kernel/ # Device-side code (Ascend C kernel on AI Core)
│ └── <op_name>_kernel.cpp │ └── <op_name>_kernel.cpp
├── include/ ├── include/
│ └── sgl_kenel_npu_ops.h # C++ interface declarations │ └── sgl_kenel_npu_ops.h # C++ interface declarations
@@ -62,7 +62,7 @@ sgl-kernel-npu/
A complete Ascend C operator consists of two parts: A complete Ascend C operator consists of two parts:
- **Device part**: Kernel code running on the NPU AICore, responsible for actual - **Device part**: Kernel code running on the NPU AI Core, responsible for actual
computation. Written using the Ascend C API. computation. Written using the Ascend C API.
- **Host part**: Code running on the CPU, responsible for parameter validation, - **Host part**: Code running on the CPU, responsible for parameter validation,
data pre-processing, tiling, and kernel launch. data pre-processing, tiling, and kernel launch.
@@ -86,7 +86,7 @@ csrc/<op_name>/
### Step 2: Write the Device-side Kernel (op_kernel) ### Step 2: Write the Device-side Kernel (op_kernel)
Device-side code runs on AICore and follows the Ascend C programming model. The Device-side code runs on AI Core and follows the Ascend C programming model. The
core structure is a class with `Init()` and `Process()` methods, plus an core structure is a class with `Init()` and `Process()` methods, plus an
`extern "C"` entry function. `extern "C"` entry function.
@@ -155,7 +155,7 @@ extern "C" __global__ __aicore__ void helloworld(
**Key points:** **Key points:**
- Class methods must be marked with `__aicore__`, indicating they run on AICore. - Class methods must be marked with `__aicore__`, indicating they run on AI Core.
- Use `AscendC::TPipe` + `AscendC::TQue` to build a pipeline that overlaps data - Use `AscendC::TPipe` + `AscendC::TQue` to build a pipeline that overlaps data
movement and computation. movement and computation.
- The entry function must be declared `extern "C" __global__ __aicore__`. The - The entry function must be declared `extern "C" __global__ __aicore__`. The
@@ -267,6 +267,7 @@ This is a common workaround for corporate proxy environments. If it does not res
If you cannot obtain the CA certificate, download datasets manually as shown in [Download Dataset Error](#download-dataset-error) below. If you cannot obtain the CA certificate, download datasets manually as shown in [Download Dataset Error](#download-dataset-error) below.
### EvalScope Request Retry Timeout ### EvalScope Request Retry Timeout
If EvalScope keeps retrying requests with errors like: If EvalScope keeps retrying requests with errors like:
``` ```
2026-06-22 03:09:03 - evalscope - WARNING: Attempt 4 / 5 failed: ....... Retrying... 2026-06-22 03:09:03 - evalscope - WARNING: Attempt 4 / 5 failed: ....... Retrying...
@@ -42,13 +42,13 @@ Thread 0x0000fff873f6f120 (most recent call first):
(SGLangEngine pid=3872176) [ERROR] 2026-04-14-12:14:41 (PID:3874122, Device:0, RankID:-1) ERR00100 PTA call acl api failed (SGLangEngine pid=3872176) [ERROR] 2026-04-14-12:14:41 (PID:3874122, Device:0, RankID:-1) ERR00100 PTA call acl api failed
(SGLangEngine pid=3872176) [Error]: An internal error occurs in the runtime module on the host. (SGLangEngine pid=3872176) [Error]: An internal error occurs in the runtime module on the host.
(SGLangEngine pid=3872176) Rectify the fault based on the error information in the ascend log. (SGLangEngine pid=3872176) Rectify the fault based on the error information in the ascend log.
(SGLangEngine pid=3872176) [PID: 3874122] 2026-04-14-12:14:41.897.548 AclNN_Runtime_Error(EZ9903): aclrtLaunchKerneWithHostArgs failed, return: 507000 (SGLangEngine pid=3872176) [PID: 3874122] 2026-04-14-12:14:41.897.548 AclNN_Runtime_Error(EZ9903): aclrtLaunchKernelWithHostArgs failed, return: 507000
(SGLangEngine pid=3872176) Solution: In this scenario, collect the plog when the fault occurs and locate the fault based on the plog. (SGLangEngine pid=3872176) Solution: In this scenario, collect the plog when the fault occurs and locate the fault based on the plog.
(SGLangEngine pid=3872176) TraceBack (most recent call last): (SGLangEngine pid=3872176) TraceBack (most recent call last):
(SGLangEngine pid=3872176) Check kernel task failed, stream_id=2028, task_id=48, retCode=0x7080005.[FUNC:LaunchKernel][FILE:context.cc][LINE:1585] (SGLangEngine pid=3872176) Check kernel task failed, stream_id=2028, task_id=48, retCode=0x7080005.[FUNC:LaunchKernel][FILE:context.cc][LINE:1585]
(SGLangEngine pid=3872176) rtsLaunchKernelWithHostArgs execution failed, reason=kernel type error[FUNC:FuncErrorReason][FILE:error_message_manage.cc][LINE:61] (SGLangEngine pid=3872176) rtsLaunchKernelWithHostArgs execution failed, reason=kernel type error[FUNC:FuncErrorReason][FILE:error_message_manage.cc][LINE:61]
(SGLangEngine pid=3872176) rtsLaunchKernelWithHostArgs failed, runtime result = 507000.[FUNC:ReportCallError][FILE:Log_inner.cpp][LINE:148] (SGLangEngine pid=3872176) rtsLaunchKernelWithHostArgs failed, runtime result = 507000.[FUNC:ReportCallError][FILE:Log_inner.cpp][LINE:148]
(SGLangEngine pid=3872176) aclrtLaunchKerneWWithHostArgs failed, return: 507000 (SGLangEngine pid=3872176) aclrtLaunchKernelWithHostArgs failed, return: 507000
(SGLangEngine pid=3872176) Launch kernel failed. (SGLangEngine pid=3872176) Launch kernel failed.
(SGLangEngine pid=3872176) #### KernelLaunch failed: /home/850b160/cann-8.5.8/opp/built-in/op_impl/ai_core/tbe//kernel/ascend910_93/ops_legacy/fill/Fill_41dadce325bOf810d03359af2a38990b_high_performance.o (SGLangEngine pid=3872176) #### KernelLaunch failed: /home/850b160/cann-8.5.8/opp/built-in/op_impl/ai_core/tbe//kernel/ascend910_93/ops_legacy/fill/Fill_41dadce325bOf810d03359af2a38990b_high_performance.o
(SGLangEngine pid=3872176) Kernel Run failed. opType: 18, Fill (SGLangEngine pid=3872176) Kernel Run failed. opType: 18, Fill
@@ -109,7 +109,7 @@ prefill 7.log:3094:RuntimeError: ACL stream synchronize failed, error code:50703
[ERROR] RUNTIME(164629,):2026-04-14-05:34:12.381.705 davinci kernel task.cc:14241170327 GetArasInfo:[AIC INFO] aras(0 to 9) after execute:0x3fffffb9000, 0, 0, 0x12c9323ff600, 0x12c93231ee00, 0x12c93f1d8800,0x12c93f3ff600,0x12c958200000,0x100000003, 0xaaaa00000001. [ERROR] RUNTIME(164629,):2026-04-14-05:34:12.381.705 davinci kernel task.cc:14241170327 GetArasInfo:[AIC INFO] aras(0 to 9) after execute:0x3fffffb9000, 0, 0, 0x12c9323ff600, 0x12c93231ee00, 0x12c93f1d8800,0x12c93f3ff600,0x12c958200000,0x100000003, 0xaaaa00000001.
[ERROR] RUNTIME(164629,):2026-04-14-05:34:12.381.710 [davinci kernel task.cc:1427]170327 GetArgsInfo:tilingKey = 0, print 1 Times totalLen=(10*8), argsSize=80, blockDim=3 [ERROR] RUNTIME(164629,):2026-04-14-05:34:12.381.710 [davinci kernel task.cc:1427]170327 GetArgsInfo:tilingKey = 0, print 1 Times totalLen=(10*8), argsSize=80, blockDim=3
[ERROR] RUNTIME(164629,):2026-04-14-05:34:12.381.717 [davinci kernel task.cc:1468]170327 PrintErrorInfoForDavinciTask:[AIC INFO] after execute:arqs print end [ERROR] RUNTIME(164629,):2026-04-14-05:34:12.381.717 [davinci kernel task.cc:1468]170327 PrintErrorInfoForDavinciTask:[AIC INFO] after execute:arqs print end
[ERROR] RUNTIME(164629,):2026-04-14-05:34:12.381.751 davinci kernel task.cc:14981170327 PrintErrorInfoForDavinciTask:[DFX INFO]Aicore kernel execute failed, device id=13, stream id=43, report stream id=43, task id=13145, flip num=56, fault kernel_name=alloc_extend _kernel_18, fault kernel info ext=alloc_extend_kernel, program id=141, hash=14069671779787989248. [ERROR] RUNTIME(164629,):2026-04-14-05:34:12.381.751 davinci kernel task.cc:14981170327 PrintErrorInfoForDavinciTask:[DFX INFO]Aicore kernel execute failed, device id=13, stream id=43, report stream id=43, task id=13145, flip num=56, fault kernel_name=alloc_extend_kernel_18, fault kernel info ext=alloc_extend_kernel, program id=141, hash=14069671779787989248.
[ERROR] IDEDD(164629,):2026-04-14-05:34:12.381.823 [dump manager.cpp:41][tid:170327] An exception callback message is received. [ERROR] IDEDD(164629,):2026-04-14-05:34:12.381.823 [dump manager.cpp:41][tid:170327] An exception callback message is received.
[ERROR] IDEDD(164629,):2026-04-14-05:34:12.381.971 [kernel info collector.cpp:384][tid:170327] Get error register information. coreNum=0 [ERROR] IDEDD(164629,):2026-04-14-05:34:12.381.971 [kernel info collector.cpp:384][tid:170327] Get error register information. coreNum=0
[ERROR] IDEDD(164629,):2026-04-14-05:34:12.381.981 kernel info collector.cpp:4771tid:1703271 It is Non-SuperKernel. functionCount=1, qlobalCount=1 [ERROR] IDEDD(164629,):2026-04-14-05:34:12.381.981 kernel info collector.cpp:4771tid:1703271 It is Non-SuperKernel. functionCount=1, qlobalCount=1
@@ -119,7 +119,7 @@ prefill 7.log:3094:RuntimeError: ACL stream synchronize failed, error code:50703
[ERROR] RUNTIME(164629,):2026-04-14-05:34:12.383.911 [stream.cc:1335]170327 GetError:AIV Kernel happen error, retCode=0x31. [ERROR] RUNTIME(164629,):2026-04-14-05:34:12.383.911 [stream.cc:1335]170327 GetError:AIV Kernel happen error, retCode=0x31.
[ERROR] RUNTIME(164629,):2026-04-14-05:34:12.383.929 [stream.cc:1335]170327 GetError:[AIC_INFO] after execute:args print end [ERROR] RUNTIME(164629,):2026-04-14-05:34:12.383.929 [stream.cc:1335]170327 GetError:[AIC_INFO] after execute:args print end
[ERROR] RUNTIME(164629,):2026-04-14-05:34:12.383.936 stream.cc:13351170327 GetError: DFX INFO1Aicore kernel execute failed, device id=13, stream id=43, report stream id=43, task id=13145, tlip num=56, [ERROR] RUNTIME(164629,):2026-04-14-05:34:12.383.936 stream.cc:13351170327 GetError: DFX INFO1Aicore kernel execute failed, device id=13, stream id=43, report stream id=43, task id=13145, tlip num=56,
fault kernel_name=alloc_extend_kernel_18, ault kernel info ext=alloc_extend_kernel, program id=141, hash=14069671779787989248. fault kernel_name=alloc_extend_kernel_18, fault kernel info ext=alloc_extend_kernel, program id=141, hash=14069671779787989248.
[ERROR] RUNTIME(164629,):2026-04-14-05:34:12.383.943 [stream.cc:3549]170327 EnterFailureAbort:stream id=43 enter failure abort. [ERROR] RUNTIME(164629,):2026-04-14-05:34:12.383.943 [stream.cc:3549]170327 EnterFailureAbort:stream id=43 enter failure abort.
[ERROR] RUNTIME(164629,):2026-04-14-05:34:12.383.973 [stars_engine.cc:1427]170327 StarsResumeRtsa:stop scheduling in abort failure mode: stream id=43, sq id=6,sq head=801, task id=13145, taskType=66. [ERROR] RUNTIME(164629,):2026-04-14-05:34:12.383.973 [stars_engine.cc:1427]170327 StarsResumeRtsa:stop scheduling in abort failure mode: stream id=43, sq id=6,sq head=801, task id=13145, taskType=66.
[ERROR] RUNTIME(164629,):2026-04-14-05:34:12.383.972 [stream.cc:1463]164629 SynchronizeExecutedTask:context is abort, status=0x715005e. [ERROR] RUNTIME(164629,):2026-04-14-05:34:12.383.972 [stream.cc:1463]164629 SynchronizeExecutedTask:context is abort, status=0x715005e.
@@ -33,9 +33,9 @@ pip install -e .
Current SGLang-MindSpore supports Qwen3 and DeepSeek V3/R1 models. This doc uses Qwen3-8B as an example. Current SGLang-MindSpore supports Qwen3 and DeepSeek V3/R1 models. This doc uses Qwen3-8B as an example.
### Offline infer ### Offline Inference
Use the following script for offline infer: Use the following script for offline inference:
<CodeGroup> <CodeGroup>
```python Offline Inference ```python Offline Inference
@@ -107,7 +107,7 @@ python3 -m sglang.launch_server \
## Troubleshooting ## Troubleshooting
#### Debug Mode ### Debug Mode
Enable sglang debug logging by log-level argument. Enable sglang debug logging by log-level argument.
@@ -132,7 +132,7 @@ export GLOG_v=0 # DEBUG
``` ```
</CodeGroup> </CodeGroup>
#### Explicitly select devices ### Explicitly select devices
Use the following environment variable to explicitly select the devices to use. Use the following environment variable to explicitly select the devices to use.
@@ -142,7 +142,7 @@ export ASCEND_RT_VISIBLE_DEVICES=4,5,6,7 # to set device
``` ```
</CodeGroup> </CodeGroup>
#### Some communication environment issues ### Some communication environment issues
In case of some environment with special communication environment, users need set some environment variables. In case of some environment with special communication environment, users need set some environment variables.
@@ -152,7 +152,7 @@ export MS_ENABLE_LCCL=off # current not support LCCL communication mode in SGLan
``` ```
</CodeGroup> </CodeGroup>
#### Some dependencies of protobuf ### Some dependencies of protobuf
In case of some environment with special protobuf version, users need set some environment variables to avoid binary version mismatch. In case of some environment with special protobuf version, users need set some environment variables to avoid binary version mismatch.
@@ -1,7 +1,7 @@
--- ---
title: "Ring-SP Performance (Wan2.1-T2V-1.3B)" title: "Ring-SP Performance (Wan2.1-T2V-1.3B)"
metatags: metatags:
description: "This page reports Ring-SP performance on Ascend NPU with torchnpu==2.10.0." description: "This page reports Ring-SP performance on Ascend NPU with torch_npu==2.10.0."
--- ---
This page reports Ring-SP performance on Ascend NPU with `torch_npu==2.10.0`. This page reports Ring-SP performance on Ascend NPU with `torch_npu==2.10.0`.
@@ -37,7 +37,7 @@ Current SGLang-MindSpore supports Qwen3 and DeepSeek V3/R1 models. This doc uses
Use the following script for offline inference: Use the following script for offline inference:
```python Offline Infer ```python Offline Inference
import sglang as sgl import sglang as sgl
# Initialize the engine with MindSpore backend # Initialize the engine with MindSpore backend