Frontend Heuristic Policy Enumeration API (RFC 0007 - Part 3/3) (ROCm#6757)

## Overview

This PR completes the three-part implementation of RFC 0007 by adding
**frontend C++ wrapper APIs** for heuristic policy enumeration. It
provides a user-friendly interface for applications to discover and
inspect loaded heuristic policies at runtime, enabling diagnostics,
logging, and understanding of engine selection behavior.

**PR Series:**
- **PR1:** Plugin SDK Infrastructure
(`users/cderb/rfc0007/pr1-plugin-sdk-infrastructure`)
- **PR2:** Heuristic Framework + Default Plugins
(`users/cderb/rfc0007/pr2-heuristic-framework-plugins`)
- **PR3:** Frontend Wrapper API
(`users/cderb/rfc0007/pr3-frontend-wrapper-clean`) <-- **This PR**

---

## What's New in PR3

### 1. Frontend Policy Enumeration API

**New Header:**
`frontend/include/hipdnn_frontend/HeuristicPolicyInfo.hpp`

Provides a clean C++ interface to query loaded heuristic policies:

```cpp
#include <hipdnn_frontend/HeuristicPolicyInfo.hpp>

auto [handle, err] = hipdnn_frontend::createHipdnnHandle();

// Query all loaded heuristic policies
auto [policies, err2] = hipdnn_frontend::getLoadedHeuristicPolicyInfos(*handle);

for (const auto& policy : policies) {
    std::cout << "Policy: " << policy.policyName
              << " (ID: " << policy.policyId << ")"
              << " Version: " << policy.pluginVersion
              << " API: " << policy.apiVersion << std::endl;
}
```

**Output Example:**
```
Policy: SelectionHeuristic::Config (ID: -7234567890123456789) Version: 1.0.0 API: 0.0.1
Policy: SelectionHeuristic::StaticOrdering (ID: -1234567890123456789) Version: 1.0.0 API: 0.0.1
```

### 2. Backend C API Extensions

**New Functions in `hipdnn_backend.h`:**

```c
// Get count of loaded heuristic policies
hipdnnStatus_t hipdnnGetHeuristicPolicyCount_ext(
    hipdnnHandle_t handle,
    size_t* numPolicies);

// Query policy metadata by index (two-call pattern for strings)
hipdnnStatus_t hipdnnGetHeuristicPolicyInfo_ext(
    hipdnnHandle_t handle,
    size_t index,
    int64_t* policyId,
    char* policyName,
    size_t* policyNameLen,
    char* pluginVersion,
    size_t* pluginVersionLen,
    char* apiVersion,
    size_t* apiVersionLen);
```

### 3. DeviceProperties Helper Functions

**New Files:**
- `backend/src/heuristics/DeviceProperties.hpp`
- `backend/src/heuristics/DeviceProperties.cpp`

Provides utility functions for heuristic policy plugins:

```cpp
// Query device properties from HIP
hipdnn_flatbuffers_sdk::data_objects::DevicePropertiesT
    queryDeviceProperties();

// Serialize device properties to FlatBuffer
std::vector<uint8_t>
    serializeDeviceProperties(const DevicePropertiesT& props);

// Wrap serialized buffer in plugin data structure
hipdnnPluginConstData_t
    wrapSerializedDeviceProperties(const std::vector<uint8_t>& serializedBuffer);
```

**Usage in Backend:**
```cpp
// Query and serialize device properties
auto devProps = queryDeviceProperties();
auto serialized = serializeDeviceProperties(devProps);
auto wrapper = wrapSerializedDeviceProperties(serialized);

// Pass to heuristic plugin
hipdnnHeuristicHandleSetDeviceProperties(pluginHandle, &wrapper);
```

---

## Files Changed

### New Files (5)

1. **Frontend API:**
- `frontend/include/hipdnn_frontend/HeuristicPolicyInfo.hpp` - Policy
enumeration API

2. **Backend Helpers:**
- `backend/src/heuristics/DeviceProperties.hpp` - Device properties
utilities (header)
- `backend/src/heuristics/DeviceProperties.cpp` - Device properties
utilities (impl)

3. **Tests:**
- `frontend/tests/TestHeuristicPolicyEnumeration.cpp` - Frontend API
tests (12 tests)

4. **Mock Backend:**
   - (Updates to existing mock for testing)

### Modified Files (4)

1. **Backend Wrappers:**
- `frontend/include/hipdnn_frontend/detail/BackendWrapper.hpp` - Add
policy query wrappers
- `frontend/include/hipdnn_frontend/detail/HipdnnBackendInterface.hpp` -
Add virtual methods
- `frontend/include/hipdnn_frontend/detail/IncompatibleBackend.hpp` -
Graceful degradation

2. **Test Configuration:**
   - `frontend/tests/CMakeLists.txt` - Register new test file

3. **Mock Backend:**
- `frontend/tests/fake_backend/MockHipdnnBackend.hpp` - Implement mock
policy queries

**Total: 9 files changed (+583 lines, -83 lines)**

---

## Testing

### Frontend Tests (12 tests)

**File:** `frontend/tests/TestHeuristicPolicyEnumeration.cpp`

12 tests covering basic enumeration, metadata validation, default policy
presence, and integration scenarios.

### Integration with PR2 Tests

PR3 builds on PR2's comprehensive backend tests (178+ tests):
- [x] Policy loading and validation
- [x] Outer loop orchestration
- [x] Device properties serialization
- [x] Default policies (Config, StaticOrdering)

**Combined Test Coverage: 190+ tests**

---

## RFC 0007 Compliance

### Section 16: Frontend API Flow

**RFC Requirements:**
> "Frontend should expose an enumeration API similar to engine config
enumeration:
>
> - `getLoadedHeuristicPolicyInfos(hipdnnHandle_t)` returns policy
metadata
> - Policy ID (int64_t)
> - Policy name (UTF-8 string)
> - Plugin version
> - API version"

**Implementation Status:**
- [x] Frontend function `getLoadedHeuristicPolicyInfos()` implemented
- [x] Backend C API `hipdnnGetHeuristicPolicyCount_ext()` implemented
- [x] Backend C API `hipdnnGetHeuristicPolicyInfo_ext()` implemented
- [x] `HeuristicPolicyInfo` struct with all required fields
- [x] Two-call pattern for string sizing (matches engine config pattern)
- [x] Error handling via frontend `Error` type
- [x] Works before graph creation (as required)

---

## Design Decisions

### DeviceProperties Helper Functions

While RFC 0007 specifies device properties serialization for plugins,
PR3 adds reusable helper functions to avoid code duplication:

```cpp
queryDeviceProperties()          // Query from HIP
serializeDeviceProperties()      // Convert to FlatBuffer
wrapSerializedDeviceProperties() // Wrap in C API struct
```

These provide a reference implementation for custom plugin developers
and maintain FlatBuffer encapsulation without exposing POD structs
across the ABI boundary.

---

## Integration with PR1 & PR2

### From PR1 (Plugin SDK):
- [x] Uses `HeuristicsPluginApi.h` C ABI definitions
- [x] Relies on two-tier plugin architecture (handle + descriptor)

### From PR2 (Heuristic Framework):
- [x] Queries
`HeuristicPluginResourceManager::getHeuristicPolicyInfos()`
- [x] Enumerates policies loaded via `HeuristicPluginManager`
- [x] Accesses default policies (Config, StaticOrdering)

### New in PR3:
- [x] Frontend C++ wrapper API
- [x] Backend C API extensions (`hipdnnGetHeuristicPolicy*_ext`)
- [x] DeviceProperties helper utilities
- [x] Frontend integration tests

---

## Summary

PR3 completes RFC 0007 by adding frontend policy enumeration and device
properties helpers. **Combined with PR1 and PR2, RFC 0007 is now fully
implemented.**

---

---------

Co-authored-by: Claude Opus 4.6 <noreply@anthropic.com>
Co-authored-by: Jason Campbell <jascampb@amd.com>
Co-authored-by: Claude Filter Strategist <claude@anthropic.com>

Author: 3 people

projects/hipdnn/backend/src/CMakeLists.txt

@@ -55,6 +55,7 @@ add_library(
     logging/GraphLogger.cpp
     logging/Logging.cpp
     FlatbufferUtilities.cpp
+    heuristics/DeviceProperties.cpp
     heuristics/BuiltInHeuristics.cpp
     heuristics/SelectionHeuristic.cpp
     heuristics/config/ConfigBuiltIn.cpp

projects/hipdnn/backend/src/descriptors/EngineHeuristicDescriptor.cpp

@@ -14,13 +14,11 @@
 #include "utilities/EngineOrdering.hpp"
 
 // Heuristics framework
+#include "heuristics/DeviceProperties.hpp"
 #include "heuristics/SelectionHeuristic.hpp"
 #include "logging/Logging.hpp"
 #include "plugin/HeuristicPlugin.hpp"
 #include "plugin/HeuristicPluginResourceManager.hpp"
-#include <flatbuffers/flatbuffers.h>
-#include <hip/hip_runtime.h>
-#include <hipdnn_flatbuffers_sdk/data_objects/device_properties_generated.h>
 
 #include <hipdnn_data_sdk/utilities/EngineNames.hpp>
 #include <hipdnn_data_sdk/utilities/PlatformUtils.hpp>
@@ -207,43 +205,11 @@ void EngineHeuristicDescriptor::finalize()
         return;
     }
 
-    // Query and serialize device properties for the device the handle's stream
-    // is bound to.
-    int deviceId = 0;
-    auto status = hipStreamGetDevice(handle->getStream(), &deviceId);
-    if(status != hipSuccess)
-    {
-        throw HipdnnException(HIPDNN_STATUS_INTERNAL_ERROR,
-                              "Failed to get device from handle's stream");
-    }
-
-    hipDeviceProp_t hipProps;
-    status = hipGetDeviceProperties(&hipProps, deviceId);
-    if(status != hipSuccess)
-    {
-        throw HipdnnException(HIPDNN_STATUS_INTERNAL_ERROR, "Failed to get device properties");
-    }
-
-    // Create DevicePropertiesT from HIP device properties
-    hipdnn_flatbuffers_sdk::data_objects::DevicePropertiesT devProps;
-    devProps.device_id = deviceId;
-    devProps.multi_processor_count = hipProps.multiProcessorCount;
-    devProps.total_global_mem = hipProps.totalGlobalMem;
-    devProps.architecture_name = hipProps.gcnArchName;
-
-    // Serialize DevicePropertiesT using FlatBuffers
-    flatbuffers::FlatBufferBuilder builder(256);
-    auto offset = hipdnn_flatbuffers_sdk::data_objects::DeviceProperties::Pack(builder, &devProps);
-    builder.Finish(offset, "HDDP");
-
-    // Copy serialized data to persistent storage
-    std::vector<uint8_t> devicePropsSerialized(builder.GetBufferPointer(),
-                                               builder.GetBufferPointer() + builder.GetSize());
-
-    // Wrap serialized buffer in hipdnnPluginConstData_t
-    hipdnnPluginConstData_t devicePropsWrapper;
-    devicePropsWrapper.ptr = devicePropsSerialized.data();
-    devicePropsWrapper.size = devicePropsSerialized.size();
+    // devicePropsSerialized must outlive devicePropsWrapper — the wrapper aliases its storage.
+    const auto devProps = heuristics::queryDeviceProperties(handle);
+    const auto devicePropsSerialized = heuristics::serializeDeviceProperties(devProps);
+    const hipdnnPluginConstData_t devicePropsWrapper
+        = heuristics::wrapSerializedDeviceProperties(devicePropsSerialized);
 
     // Get serialized graph from GraphDescriptor
     const hipdnnPluginConstData_t serializedGraph = _graph->getSerializedGraph();

projects/hipdnn/backend/src/heuristics/DeviceProperties.cpp

@@ -0,0 +1,64 @@
+// Copyright © Advanced Micro Devices, Inc., or its affiliates.
+// SPDX-License-Identifier:  MIT
+
+#include "DeviceProperties.hpp"
+
+#include <flatbuffers/flatbuffers.h>
+#include <hip/hip_runtime.h>
+
+#include <string>
+
+#include "HipdnnException.hpp"
+#include "handle/Handle.hpp"
+
+namespace hipdnn_backend::heuristics
+{
+
+hipdnn_flatbuffers_sdk::data_objects::DevicePropertiesT queryDeviceProperties(int deviceId)
+{
+    hipDeviceProp_t hipProps;
+    auto status = hipGetDeviceProperties(&hipProps, deviceId);
+    if(status != hipSuccess)
+    {
+        throw HipdnnException(HIPDNN_STATUS_INTERNAL_ERROR,
+                              "Failed to get properties for device " + std::to_string(deviceId)
+                                  + ": " + hipGetErrorString(status));
+    }
+
+    hipdnn_flatbuffers_sdk::data_objects::DevicePropertiesT devProps;
+    devProps.device_id = deviceId;
+    devProps.multi_processor_count = hipProps.multiProcessorCount;
+    devProps.total_global_mem = hipProps.totalGlobalMem;
+    devProps.architecture_name = hipProps.gcnArchName;
+
+    return devProps;
+}
+
+hipdnn_flatbuffers_sdk::data_objects::DevicePropertiesT queryDeviceProperties(hipdnnHandle_t handle)
+{
+    if(handle == nullptr)
+    {
+        throw HipdnnException(HIPDNN_STATUS_BAD_PARAM, "queryDeviceProperties: handle is null");
+    }
+
+    int deviceId = 0;
+    auto status = hipStreamGetDevice(handle->getStream(), &deviceId);
+    if(status != hipSuccess)
+    {
+        throw HipdnnException(HIPDNN_STATUS_INTERNAL_ERROR,
+                              std::string{"Failed to get device from handle's stream: "}
+                                  + hipGetErrorString(status));
+    }
+    return queryDeviceProperties(deviceId);
+}
+
+std::vector<uint8_t>
+    serializeDeviceProperties(const hipdnn_flatbuffers_sdk::data_objects::DevicePropertiesT& props)
+{
+    flatbuffers::FlatBufferBuilder builder(256);
+    auto offset = hipdnn_flatbuffers_sdk::data_objects::DeviceProperties::Pack(builder, &props);
+    builder.Finish(offset, "HDDP");
+    return {builder.GetBufferPointer(), builder.GetBufferPointer() + builder.GetSize()};
+}
+
+} // namespace hipdnn_backend::heuristics

projects/hipdnn/backend/src/heuristics/DeviceProperties.hpp

@@ -0,0 +1,88 @@
+// Copyright © Advanced Micro Devices, Inc., or its affiliates.
+// SPDX-License-Identifier:  MIT
+
+#pragma once
+
+#include <cstddef>
+#include <vector>
+
+#include <hipdnn_flatbuffers_sdk/data_objects/device_properties_generated.h>
+#include <hipdnn_plugin_sdk/PluginApiDataTypes.h>
+
+struct hipdnnHandle;
+using hipdnnHandle_t = struct hipdnnHandle*;
+
+namespace hipdnn_backend::heuristics
+{
+
+/**
+ * @brief Query device properties for an explicit device id.
+ *
+ * Calls hipGetDeviceProperties() for @p deviceId and populates a
+ * DevicePropertiesT structure. Use this overload when the caller already
+ * knows which device to query (e.g. resolved from a handle's stream).
+ *
+ * RFC 0007 Reference: Section 6.2
+ *
+ * @param deviceId HIP device id to query.
+ * @return DevicePropertiesT populated from HIP device properties.
+ * @throws HipdnnException if the HIP call fails.
+ */
+hipdnn_flatbuffers_sdk::data_objects::DevicePropertiesT queryDeviceProperties(int deviceId);
+
+/**
+ * @brief Query device properties for the device bound to a handle's stream.
+ *
+ * Resolves the device id from @p handle's stream via hipStreamGetDevice and
+ * delegates to the explicit-id overload. This is the canonical acquisition
+ * path inside the backend: device facts must follow the handle's stream, not
+ * whatever device happens to be current on the calling thread.
+ *
+ * RFC 0007 Reference: Section 6.2
+ *
+ * @param handle Handle whose stream identifies the target device.
+ * @return DevicePropertiesT populated from HIP device properties.
+ * @throws HipdnnException if @p handle is null or if any HIP call fails.
+ */
+hipdnn_flatbuffers_sdk::data_objects::DevicePropertiesT
+    queryDeviceProperties(hipdnnHandle_t handle);
+
+/**
+ * @brief Serialize DevicePropertiesT to FlatBuffer format.
+ *
+ * Builds a FlatBuffer-serialized representation of the device properties
+ * using the Pack method and returns the bytes in a freshly owned vector
+ * (the caller owns the storage; the internal FlatBufferBuilder is released
+ * before return).
+ *
+ * The returned vector must remain alive while any hipdnnPluginConstData_t
+ * wrapper produced by wrapSerializedDeviceProperties is in use, because the
+ * wrapper aliases the vector's buffer.
+ *
+ * RFC 0007 Reference: Section 13.2
+ *
+ * @param props Device properties to serialize.
+ * @return Vector owning the serialized FlatBuffer bytes.
+ */
+std::vector<uint8_t>
+    serializeDeviceProperties(const hipdnn_flatbuffers_sdk::data_objects::DevicePropertiesT& props);
+
+/**
+ * @brief Wrap serialized device properties in hipdnnPluginConstData_t.
+ *
+ * Creates a hipdnnPluginConstData_t wrapper pointing to the serialized buffer.
+ * The buffer must remain valid while the wrapper is in use.
+ *
+ * @param serializedBuffer Reference to the serialized buffer (must outlive the wrapper).
+ * @return hipdnnPluginConstData_t wrapper pointing to the buffer.
+ */
+inline hipdnnPluginConstData_t
+    wrapSerializedDeviceProperties(const std::vector<uint8_t>& serializedBuffer)
+{
+    hipdnnPluginConstData_t wrapper;
+    wrapper.ptr = serializedBuffer.data();
+    wrapper.size = serializedBuffer.size();
+    return wrapper;
+}
+
+} // namespace hipdnn_backend::heuristics

projects/hipdnn/backend/tests/CMakeLists.txt

@@ -108,9 +108,10 @@ add_executable(
     TestHeuristicPluginManagerValidationPaths.cpp
     TestSelectionHeuristic.cpp
     TestHeuristicPluginResourceManager.cpp
-    heuristics/TestStaticOrderingBuiltIn.cpp
-    heuristics/TestEngineOverrideConfig.cpp
     heuristics/TestConfigBuiltIn.cpp
+    heuristics/TestDeviceProperties.cpp
+    heuristics/TestEngineOverrideConfig.cpp
+    heuristics/TestStaticOrderingBuiltIn.cpp
     TestUserLoggingApis.cpp
     TestPlatformUtils.cpp
     TestEnginePluginOverride.cpp

projects/hipdnn/backend/tests/IntegrationHeuristicPlugin.cpp

@@ -20,6 +20,7 @@
 #include "HipdnnException.hpp"
 #include "PlatformUtils.hpp"
 #include "TestPluginConstants.hpp"
+#include "heuristics/DeviceProperties.hpp"
 #include "plugin/HeuristicPlugin.hpp"
 #include "plugin/HeuristicPluginManager.hpp"
 #include "plugin/HeuristicPluginResourceManager.hpp"
@@ -31,8 +32,6 @@
 #include <hipdnn_data_sdk/utilities/ScopedResource.hpp>
 #include <hipdnn_flatbuffers_sdk/data_objects/device_properties_generated.h>
 
-#include <flatbuffers/flatbuffers.h>
-
 #include <gtest/gtest.h>
 #include <string_view>
 
@@ -45,15 +44,8 @@ namespace
 // Note: TEST_GOOD_HEURISTIC_PLUGIN_NAME, TEST_INCOMPLETE_HEURISTIC_API_PLUGIN_NAME,
 // and TEST_NO_OPTIONAL_HEURISTIC_PLUGIN_NAME are defined as macros in CMakeLists.txt
 
-// Helper to serialize DevicePropertiesT using FlatBuffers Pack
-std::vector<uint8_t>
-    serializeDeviceProperties(const hipdnn_flatbuffers_sdk::data_objects::DevicePropertiesT& props)
-{
-    flatbuffers::FlatBufferBuilder builder(256);
-    auto offset = hipdnn_flatbuffers_sdk::data_objects::DeviceProperties::Pack(builder, &props);
-    builder.Finish(offset, "HDDP");
-    return {builder.GetBufferPointer(), builder.GetBufferPointer() + builder.GetSize()};
-}
+using hipdnn_backend::heuristics::serializeDeviceProperties;
+using hipdnn_backend::heuristics::wrapSerializedDeviceProperties;
 
 // Wrapper class to access protected constructor
 class TestableHeuristicPlugin : public HeuristicPlugin
@@ -172,10 +164,8 @@ TEST_F(IntegrationHeuristicPlugin, SetDevicePropertiesOnHandle)
     props.architecture_name = "gfx90a";
 
     // Serialize
-    auto serialized = serializeDeviceProperties(props);
-    hipdnnPluginConstData_t devicePropsData;
-    devicePropsData.ptr = serialized.data();
-    devicePropsData.size = serialized.size();
+    const auto serialized = serializeDeviceProperties(props);
+    const hipdnnPluginConstData_t devicePropsData = wrapSerializedDeviceProperties(serialized);
 
     // Set on handle (should not throw)
     EXPECT_NO_THROW(plugin->setDeviceProperties(handle, &devicePropsData));
@@ -193,10 +183,8 @@ TEST_F(IntegrationHeuristicPlugin, SetDevicePropertiesOnAllHandles)
     props.total_global_mem = 16ULL * 1024 * 1024 * 1024;
     props.architecture_name = "gfx90a";
 
-    auto serialized = serializeDeviceProperties(props);
-    hipdnnPluginConstData_t devicePropsData;
-    devicePropsData.ptr = serialized.data();
-    devicePropsData.size = serialized.size();
+    const auto serialized = serializeDeviceProperties(props);
+    const hipdnnPluginConstData_t devicePropsData = wrapSerializedDeviceProperties(serialized);
 
     // Set on all handles via resource manager
     EXPECT_NO_THROW(rm->setDevicePropertiesOnAllHandles(&devicePropsData));
@@ -226,10 +214,8 @@ TEST_F(IntegrationHeuristicPlugin, CompleteWorkflowWithDevicePropertiesAndFinali
     props.total_global_mem = 16ULL * 1024 * 1024 * 1024;
     props.architecture_name = "gfx90a";
 
-    auto serialized = serializeDeviceProperties(props);
-    hipdnnPluginConstData_t devicePropsData;
-    devicePropsData.ptr = serialized.data();
-    devicePropsData.size = serialized.size();
+    const auto serialized = serializeDeviceProperties(props);
+    const hipdnnPluginConstData_t devicePropsData = wrapSerializedDeviceProperties(serialized);
     plugin->setDeviceProperties(handle, &devicePropsData);
 
     // Create policy descriptor (RAII so destroy runs even on ASSERT_* abort)
@@ -463,10 +449,8 @@ TEST_F(IntegrationHeuristicPlugin, SetDevicePropertiesWithNoPluginsLoaded)
     props.total_global_mem = 16ULL * 1024 * 1024 * 1024;
     props.architecture_name = "gfx90a";
 
-    auto serialized = serializeDeviceProperties(props);
-    hipdnnPluginConstData_t devicePropsData;
-    devicePropsData.ptr = serialized.data();
-    devicePropsData.size = serialized.size();
+    const auto serialized = serializeDeviceProperties(props);
+    const hipdnnPluginConstData_t devicePropsData = wrapSerializedDeviceProperties(serialized);
 
     // Should not throw when no plugins loaded
     EXPECT_NO_THROW(rm->setDevicePropertiesOnAllHandles(&devicePropsData));

projects/hipdnn/backend/tests/descriptors/TestEngineHeuristicDescriptor.cpp

@@ -81,7 +81,7 @@ class TestEngineHeuristicDescriptor : public ::testing::Test
 
     void setHeuristicMode() const
     {
-        hipdnnBackendHeurMode_t mode = HIPDNN_HEUR_MODE_FALLBACK;
+        const hipdnnBackendHeurMode_t mode = HIPDNN_HEUR_MODE_FALLBACK;
         ASSERT_NO_THROW(getEngineHeuristicDescriptor()->setAttribute(
             HIPDNN_ATTR_ENGINEHEUR_MODE, HIPDNN_TYPE_HEUR_MODE, 1, &mode));
     }