[DOCS] RDNA / Navi specific doc

CMakeLists.txt

@@ -744,3 +744,11 @@ rocm_create_package(
     LDCONFIG
     HEADER_ONLY
 )
+
+# Add documentation support
+option(BUILD_DOCS "Build documentation" ON)
+
+if(BUILD_DOCS)
+    include(cmake/SetupDocs.cmake)
+    setup_documentation()
+endif()

cmake/SetupDocs.cmake

@@ -0,0 +1,109 @@
+function(setup_documentation)
+    # Find Python executable
+    find_package(Python3 REQUIRED COMPONENTS Interpreter)
+
+    # Check for existing virtual environment in ~/python-env
+    set(HOME_VENV "$ENV{HOME}/python-env/venv-1")
+    set(BUILD_VENV "${CMAKE_BINARY_DIR}/python-env")
+
+    if(EXISTS "${HOME_VENV}/bin/python3")
+        set(PYTHON_VENV "${HOME_VENV}")
+        set(PYTHON_EXECUTABLE "${HOME_VENV}/bin/python3")
+        set(VENV_IS_NEW FALSE)
+        message(STATUS "Using existing Python virtual environment: ${HOME_VENV}")
+    else()
+        set(PYTHON_VENV "${BUILD_VENV}")
+        set(PYTHON_EXECUTABLE "${BUILD_VENV}/bin/python3")
+        set(VENV_IS_NEW TRUE)
+        message(STATUS "Will create Python virtual environment: ${BUILD_VENV}")
+    endif()
+
+    # Create virtual environment if it doesn't exist
+    if(NOT EXISTS "${PYTHON_EXECUTABLE}")
+        message(STATUS "Creating Python virtual environment...")
+        execute_process(
+            COMMAND ${Python3_EXECUTABLE} -m venv ${PYTHON_VENV}
+            RESULT_VARIABLE VENV_RESULT
+        )
+        if(NOT VENV_RESULT EQUAL 0)
+            message(FATAL_ERROR "Failed to create Python virtual environment")
+        endif()
+        set(VENV_IS_NEW TRUE)
+    endif()
+
+    # Set requirements file path
+    set(REQUIREMENTS_FILE "${CMAKE_SOURCE_DIR}/docs/sphinx/requirements.txt")
+
+    # Only install packages if we created a new virtual environment
+    if(VENV_IS_NEW)
+        message(STATUS "Installing Python documentation requirements in new virtual environment...")
+        execute_process(
+            COMMAND ${PYTHON_EXECUTABLE} -m pip install --upgrade pip
+            RESULT_VARIABLE PIP_RESULT
+        )
+        if(NOT PIP_RESULT EQUAL 0)
+            message(FATAL_ERROR "Failed to upgrade pip")
+        endif()
+
+        execute_process(
+            COMMAND ${PYTHON_EXECUTABLE} -m pip install -r ${REQUIREMENTS_FILE}
+            RESULT_VARIABLE PIP_RESULT
+        )
+        if(NOT PIP_RESULT EQUAL 0)
+            message(FATAL_ERROR "Failed to install Python documentation requirements")
+        endif()
+
+        message(STATUS "Python documentation requirements installed successfully")
+    endif()
+
+    # Create dummy target (no build-time package installation needed)
+    add_custom_target(docs_venv
+        COMMENT "Python virtual environment ready"
+    )
+
+    # Set up Doxygen
+    find_package(Doxygen REQUIRED)
+
+    set(DOXYGEN_OUTPUT_DIR "${CMAKE_BINARY_DIR}/docs/doxygen")
+    set(DOXYGEN_CONFIG_FILE "${CMAKE_SOURCE_DIR}/docs/doxygen/Doxyfile")
+
+    add_custom_command(
+        OUTPUT "${DOXYGEN_OUTPUT_DIR}/xml/index.xml"
+        COMMAND ${CMAKE_COMMAND} -E make_directory ${DOXYGEN_OUTPUT_DIR}
+        COMMAND ${DOXYGEN_EXECUTABLE} ${DOXYGEN_CONFIG_FILE}
+        WORKING_DIRECTORY ${CMAKE_SOURCE_DIR}/docs/doxygen
+        DEPENDS ${DOXYGEN_CONFIG_FILE}
+        COMMENT "Building Doxygen XML documentation"
+    )
+
+    add_custom_target(docs_doxygen
+        DEPENDS "${DOXYGEN_OUTPUT_DIR}/xml/index.xml"
+    )
+
+    # Set up Sphinx
+    set(SPHINX_OUTPUT_DIR "${CMAKE_BINARY_DIR}/docs/html")
+    set(SPHINX_SOURCE_DIR "${CMAKE_SOURCE_DIR}/docs")
+
+    # Set number of parallel jobs for Sphinx (default to 8 if not set)
+    if(NOT CMAKE_BUILD_PARALLEL_LEVEL)
+        set(CMAKE_BUILD_PARALLEL_LEVEL 8)
+    endif()
+
+    add_custom_target(docs
+        COMMAND ${PYTHON_EXECUTABLE} -m sphinx -b html -j ${CMAKE_BUILD_PARALLEL_LEVEL} ${SPHINX_SOURCE_DIR} ${SPHINX_OUTPUT_DIR}
+        WORKING_DIRECTORY ${CMAKE_SOURCE_DIR}
+        DEPENDS docs_venv docs_doxygen
+        COMMENT "Building HTML documentation with Sphinx"
+    )
+
+    # Add clean target for docs
+    add_custom_target(docs_clean
+        COMMAND ${CMAKE_COMMAND} -E remove_directory ${CMAKE_BINARY_DIR}/docs
+        COMMENT "Cleaning documentation build files"
+    )
+
+    message(STATUS "Documentation build configured:")
+    message(STATUS "  Build command: cmake --build . --target docs")
+    message(STATUS "  Clean command: cmake --build . --target docs_clean")
+    message(STATUS "  Output directory: ${SPHINX_OUTPUT_DIR}")
+endfunction()

docs/conceptual/ck_tile_radeon/hardware/index.rst

@@ -0,0 +1,44 @@
+.. meta::
+   :description: CK Tile Hardware-Specific Documentation
+   :keywords: CDNA, GPU architecture, LDS, GEMM, CK, Composable Kernel
+
+.. _ck_tile_hardware:
+
+********************************************************************
+CK Tile Hardware Documentation - Radeon / Navi
+********************************************************************
+
+This section provides in-depth coverage of hardware-specific concepts and optimizations for CK Tile on AMD GPUs.
+
+Overview
+========
+
+Understanding the underlying hardware architecture is crucial for achieving optimal performance with CK Tile. This documentation covers differences between AMD CDNA and AMD RDNA across the following areas:
+
+- Instruction Set Architecture (ISA)
+- Memory hierarchy and optimization techniques
+- Practical examples of high-performance kernels
+
+Key Hardware Considerations
+===========================
+
+When using CK Tile, keep these hardware aspects in mind:
+
+Memory Hierarchy
+----------------
+
+The memory hierarchy on Radeon / Navi is similar to the memory hiararchy on CDNA / Instinct.  It is organised into Global Memory, L2/Infinity Cache, LDS, Registers.
+
+[TODO: table with sizes at all levels for MI200, MI300, Navi31, Navi48 ]
+
+Compute Resources
+------------------
+
+1. **Wavefront Execution**: multiple threads in lockstep as on CNDA / Instinct
+
+   - 32 threads (in contrast to 64 on CDNA / Instinct)
+
+2. **Matrix Units**: specialized Matrix Multiple Units as on CDNA / Instinct
+
+   - [TODO - gfx11 / Navi 3 WMMA specs]
+   - [TODO - gfx12 / Navi 4 WMMA specs]

docs/conceptual/ck_tile_radeon/index.rst

@@ -0,0 +1,17 @@
+
+
+.. _ck_tile_radeon_conceptual:
+
+CK Tile Conceptual Documentation
+================================
+
+Welcome to the conceptual documentation for CK Tile, the core abstraction layer of Composable Kernel that enables efficient GPU programming through compile-time coordinate transformations and tile-based data distribution.
+
+Documentation Structure
+-----------------------
+
+.. toctree::
+   :maxdepth: 2
+   :caption: CK Tile Concepts - Radeon / Navi
+
+   hardware/index