Check Macros#
Defines
-
TRT_EDGELLM_STRINGIFY_IMPL(x) #x#
-
TRT_EDGELLM_STRINGIFY(x) TRT_EDGELLM_STRINGIFY_IMPL(x)#
-
ELLM_CHECK(cond, msg)
do \
{ \
if (!(cond)) \
{ \
throw std::runtime_error(std::string(__FILE__ ":" TRT_EDGELLM_STRINGIFY(__LINE__) ": ") + (msg)); \
} \
} while (0)
# Lazy-message precondition check.
Throws std::runtime_error if
condis false. Themsgexpression is only evaluated on failure, which matters when the message uses string concatenation (e.g. std::to_string, ostringstream) on hot paths. The thrown message is prefixed with FILE:LINE to aid debugging.Usage: ELLM_CHECK(ptr != nullptr, “ptr must not be null”); ELLM_CHECK(n > 0, “n must be positive, got “ + std::to_string(n));
-
CUDA_CHECK(stat)
do \
{ \
trt_edgellm::check::_checkCuda((stat), #stat, __FILE__, __LINE__); \
} while (0)
# Check CUDA runtime API calls.
Wraps CUDA runtime API calls and throws exception on error. Usage: CUDA_CHECK(cudaMalloc(&ptr, size));
-
CUDA_DRIVER_CHECK(stat)
do \
{ \
trt_edgellm::check::_checkCudaDriver((stat), #stat, __FILE__, __LINE__); \
} while (0)
# Check CUDA driver API calls.
Wraps CUDA driver API calls and throws exception on error. Usage: CUDA_DRIVER_CHECK(cuMemAlloc(&dptr, size));
-
namespace trt_edgellm
Argument bundles for the CuTe DSL FMHA launchers.
Each public run* entry point fills one of these once from its own arguments and members, and the launcher then spreads it across the generated descriptors. One struct per wrapper shape rather than a single superset: a superset would let a field that matters on one path (kvCacheCapacity on the dense path, tokensPerPage on the paged one) sit silently at zero on another.
These carry no generated types, so this header is safe to include from translation units that use the required FMHA-v2 runner, the optional CUTE_DSL_FMHA_BLACKWELL_ENABLED runner, or both. The descriptor-filling machinery itself lives in cuteDslTensorDescriptors.h, which stays free of any FMHA concept.
Helpers for populating the tensor descriptors emitted by the CuTe DSL C exporter.
Every AOT variant exports its own nominally distinct but layout-identical descriptor structs (
fmha_d64_Tensor_q_tensor_tvsfmha_d128_Tensor_q_tensor_t) plus acute_dsl_<variant>_wrapperentry point. There is no umbrella C type, so descriptor types are recovered here from the signature of the wrapper that consumes them: a call site names only the wrapper and the kernel module, and pairing a descriptor with the wrong variant is not expressible.The exporter (
cutlass/cute/export/c_header_generator.py) always names the membersdata,dynamic_shapesanddynamic_strides, but emits each array only when its dynamic mask is non-empty. A rank-1 descriptor therefore has nodynamic_stridesmember at all and needs makeCuSeqLenTensor() rather than the strided builders below.-
namespace check#
Functions
-
inline void check(bool condition, std::string errorMsg)#
Check condition and throw exception if false.
- Parameters:
condition – Condition to check
errorMsg – Error message to include in exception
- Throws:
std::runtime_error – If condition is false
- inline void _checkCuda(
- cudaError_t result,
- char const *const func,
- char const *const file,
- int const line
Internal helper to check CUDA runtime errors.
- Parameters:
result – CUDA error code
func – Function name string
file – Source file name
line – Source line number
- Throws:
std::runtime_error – If CUDA error occurred
- inline void _checkCudaDriver(
- CUresult result,
- char const *const func,
- char const *const file,
- int const line
Internal helper to check CUDA driver API errors.
- Parameters:
result – CUDA driver error code
func – Function name string
file – Source file name
line – Source line number
- Throws:
std::runtime_error – If CUDA driver error occurred
-
inline void check(bool condition, std::string errorMsg)#
-
namespace check#