Kernel Function

Kernel Function Definition

Ascend C allows you to customize kernel functions to extend C++. When a kernel function is executed on the AI processor, several threads are executed in parallel. Each thread has an independent register and stack to complete data compute tasks.

The following is an example of kernel function definition:

1

 __global__ void kernel_name(argument list)

Comply with the following rules when defining kernel functions:

Kernel Function Call

Functions in an operator program can be classified into three types: host functions, kernel functions (executed on the device), and device functions (except kernel functions). The following figure uses operator development based on the kernel launch project as an example to describe the calling relationships between the three types of functions.

• A host function can call other host functions, similar to function calling in general C/C++ programming. A host function can also call kernel functions using <<<...>>>.
• A kernel function can call device functions other than kernel functions.
• A device function (except the kernel function) is identified by the type qualifier __aicore__ and can call other device functions of the same type.

Figure 1 Calling relationships between the three types of functions in an operator program

The host calls a kernel function using the syntax of the kernel launch symbol <<<...>>>, as shown below:

1

kernel_name<<<numBlocks, threadsPerBlock, dynUBufSize, stream>>>(args...)