Skip to content

[Epic] Develop CK Builder as an API for CK and CK Tile within MIOpen #3574

Closed

Description

@shumway

Overview

The CK Builder provides a high-level, semantically-clear interface for constructing Composable Kernel convolution operations. This experimental feature focuses on formalizing the interface between MIOpen and CK, leveraging modern C++20 features (POD structs as non-type template parameters, concepts, and designated initializers) to simplify kernel instantiation and improve developer experience.

This project serves as a prototype for a more general builder pattern that will eventually extend to all CK and CK-Tile operations.

Motivation

The current CK interface requires deep knowledge of complex template parameters, making it difficult for library users like MIOpen to instantiate kernels correctly. The builder pattern addresses this by:

  • Providing compile-time validation through C++20 concepts and other best practices
  • Offering clear, self-documenting interfaces with designated initializers
  • Separating "what to compute" (signature) from "how to compute it" (algorithm)
  • Enabling testable, strongly-typed kernel construction with meaningful compiler error messages

Status

Development Period: October 2025 - March 2026

The initial scope of CK Builder is a narrow focus on MIOpen, to deliver all supported CK kernels (convolutions) to MIOpen using CK / CK Tile as a header-only library.

The builder currently supports forward convolution kernels with various implementations (XDL, WMMA, DL, Large Tensor variants). Support for backward data and backward weight convolutions is in progress.

Requirements

  • C++20 compiler support
  • CMake configuration: -D CK_EXPERIMENTAL_BUILDER=ON

Documentation

Success Criteria

MIOpen can instantiate all required convolution kernels using the builder interface, with compile-time validation ensuring configuration correctness.

Future Vision

Extend the builder pattern to all CK and CK-Tile operations beyond convolutions, making it the primary interface for library users.

Activity

  1. shumway commented on Jan 21, 2026

    @shumway
    CollaboratorAuthor

    CK Builder Status Update for 2026.01.21

    Current Status: Task Breakdown

    1. ✅ Builder signature and algorithm concepts with C++20 validation
    2. ✅ Factory dispatch system for all kernel variants (XDL, WMMA, DL, Large Tensor)
    3. ✅ Forward convolution support (1D, 2D, 3D) with multiple data types
    4. ✅ Reflection infrastructure for kernel introspection
    5. ✅ Testing framework with Given-When-Then pattern and GPU validation
    6. ✅ Comprehensive documentation (design, factory, testing)
    7. ✅ CKTile integration foundation (header installation, initial kernel support)
    8. 🔄 Backward weight convolution testing and validation
    9. 🔄 WMMA kernel reflection and algorithm concept refinements
    10. 🔄 CKTile kernel expansion (backward data/weight support)
    11. ⏳ MIOpen end-to-end integration (Critical - escalating this week)
    12. ⏳ Hardware awareness implementation (using ck_tile/core/config.hpp helpers to distinguish XDL vs WMMA)

    Current Work In-Flight (5 Open PRs)

    The team is actively working on completing backward convolution support and refining the builder infrastructure for both CK and CKTile kernels. Key efforts include: finalizing backward weight convolution testing (#3618), improving reference implementations for validation (#3604), and extending reflection capabilities for WMMA and backward weight kernels (#3592). Additionally, we're refactoring algorithm concepts to improve type safety and developer experience (#3567), and enhancing testing utilities to support more complex validation scenarios (#3563). These PRs represent the final pieces needed to complete the core builder functionality before focusing on MIOpen integration.

    Critical Path Items

    🔴 High Priority - This Week:

    • MIOpen Integration: Escalating end-to-end integration work to reduce risk and enable multi-developer scaling. The CK builder side is maturing, but we need to close the loop with MIOpen to validate the interface and identify integration issues.

    🟡 Medium Priority - Next 2 Weeks:

    • Hardware Awareness: Implement hardware detection using helper functions in include/ck_tile/core/config.hpp to automatically distinguish between XDL and WMMA kernels based on GPU architecture.
    • CKTile Expansion: Continue extending builder support to cover CKTile kernels alongside traditional CK kernels.

    Progress: 78 PRs merged, 5 PRs open | Target Completion: March 2026

  2. shumway commented on Jan 28, 2026

    @shumway
    CollaboratorAuthor

    CK Builder Status Update for 2026.01.28

    Week's Accomplishments (Jan 20-27)

    • ✅ Backward Weight Testing Complete (#3618): Added comprehensive ckt::run integration for backward weight algorithms across reference, CK, and CK Tile implementations. Introduced RunResult struct for better error handling and testing integration with SuccessfulRun() matcher. Reorganized testing structure into testing/conv/ directory with separate files for direction-specific implementations.

    • ✅ Reference Implementation Improvements (#3604): Replaced reference convolution implementation with proven CK implementation to improve validation reliability and reduce maintenance burden.

    • ✅ CK Tile Profiler (#3518): Added grouped convolution forward profiler for CK Tile, enabling performance analysis and optimization.

    • ✅ Transfer Concepts (#3535): Implemented convolution forward transfer concepts to improve type safety and compile-time validation.

    • ✅ Documentation Updates (#3573): Updated README files across the builder codebase to reflect recent architectural changes.

    • ✅ CI Fix (#3581): Disabled CK Builder for SLES15 in Jenkins CI to address compatibility issues.

    Current Work In-Flight (1 Open PR)

    • 🔄 Validation Integration (#3649): Integrating builder/testing/validation.hpp with CK's gpu_verification.hpp to reduce code duplication. This PR introduces an iterator adapter for multi-dimensional tensor validation, cleans up the verification infrastructure with abstracted kernel launches and block reductions, and improves NaN handling in test validation. Includes updated documentation for ckt::validate().

    Critical Path Items

    🔴 High Priority - This Week:

    • MIOpen Integration: This is critical for validating the builder interface in production and identifying any remaining integration issues.
    • Validation Infrastructure: Complete the integration of validation utilities (#3649) to establish a unified testing framework across CK and CK Builder.

    🟡 Medium Priority - Next 2 Weeks:

    • Hardware Awareness: Implement hardware detection using helper functions in include/ck_tile/core/config.hpp to automatically distinguish between XDL and WMMA kernels based on GPU architecture.
    • Backward Data Support: Extend builder support to cover backward data convolutions, completing the full convolution operation suite.

    Next Week's Focus

    1. Complete validation infrastructure integration (#3649)
    2. Complete end-to-end builder integration in MIOpen.
    3. Start hardware awareness implementation for automatic XDL/WMMA selection
    4. Begin backward data convolution support

    Progress: 84 PRs merged, 1 PR open | Target Completion: March 2026

  3. assistant-librarian commented on Feb 3, 2026

    @assistant-librarian
    Contributor

    This issue has been migrated to: ROCm/rocm-libraries#4231

  4. ammallya commented on Feb 3, 2026

    @ammallya

    Imported to ROCm/rocm-libraries

Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment

Metadata

Metadata

Assignees

No one assigned

    Labels

    No labels
    No labels

    Type

    No type

    Projects

    No projects

      Milestone

      No milestone

      Relationships

      None yet

      Development

      No branches or pull requests

      Issue actions