API Reference

Complete API documentation for Helion.

Kernel Creation and Control

Everything you need to create and configure Helion kernels using the helion.kernel() decorator:

• Kernel
  • Kernel
  • Overview
  • Creation and Usage
  • BoundKernel
  • Advanced Usage
  • Caching and Specialization
  • Settings vs Config in Kernel Creation
  • See Also
• Config
  • Config
  • Overview
  • Configuration Parameters
  • Usage Examples
  • See Also
• Settings
  • Settings
  • Overview
  • Configuration Examples
  • Settings Reference
  • Functions
  • See Also

Language Constructs

The helion.language module contains DSL constructs for authoring kernels:

• Language Module
  • Loop Constructs
  • Memory Operations
  • Tensor Creation
  • Tunable Parameters
  • Tile Operations
  • View Operations
  • Reduction Operations
  • Scan Operations
  • Synchronization
  • Utilities
  • Constexpr Operations

Debugging and Utilities

• Exceptions
  • Overview
  • Exception Hierarchy
  • Kernel Context Errors
  • Loop and Control Flow Errors
  • Tile and Indexing Errors
  • Assignment and Variable Errors
  • Type and Inference Errors
  • Configuration Errors
  • Tunable Parameter Errors
  • Language and Syntax Errors
  • Grid and Execution Errors
  • Compilation and Runtime Errors
  • Warning Classes
  • See Also

Advanced Topics

• Autotuner Module
  • Configuration Classes
  • Search Algorithms
• Runtime Module
  • Key Classes
  • Utility Functions

Quick Reference

Main Functions

kernel	Decorator to create a Kernel object from a Python function.
Config
Settings	Settings can be passed to hl.kernel as kwargs and control the behavior of the compilation process.

Language Functions

tile	Break up an iteration space defined by a size or sequence of sizes into tiles.
grid	Iterate over individual indices of the given iteration space.
load	Load a value from a tensor using a list of indices.
store	Store a value to a tensor using a list of indices.
atomic_add	Atomically add a value to a target tensor.
zeros	Return a device-tensor filled with zeros.
full	Create a device-tensor filled with a specified value.
arange	Same as torch.arange(), but defaults to same device as the current kernel.
reduce	Applies a reduction operation along a specified dimension or all dimensions.
associative_scan	Applies an associative scan operation along a specified dimension.
register_block_size	Explicitly register a block size that should be autotuned and can be used for allocations and inside hl.tile(..., block_size=...).
register_reduction_dim	Explicitly register a reduction dimension that should be used for reduction operations.
register_tunable	Register a tunable parameter for autotuning.
constexpr	alias of ConstExpr
specialize	Turn a dynamic shape into a compile-time constant.