refactor: add private property logging in test and kernel class

Signed-off-by: Zbigniew Zdanowicz <zbigniew.zdanowicz@intel.com>
This commit is contained in:
Zbigniew Zdanowicz
2025-01-02 10:17:50 +00:00
committed by Compute-Runtime-Automation
parent 165c294590
commit a744aa07bc
2 changed files with 8 additions and 2 deletions

View File

@@ -1,5 +1,5 @@
/*
* Copyright (C) 2020-2024 Intel Corporation
* Copyright (C) 2020-2025 Intel Corporation
*
* SPDX-License-Identifier: MIT
*
@@ -1297,6 +1297,8 @@ void ModuleImp::checkIfPrivateMemoryPerDispatchIsNeeded() {
auto globalMemorySize = neoDevice->getRootDevice()->getGlobalMemorySize(static_cast<uint32_t>(deviceBitfield.to_ulong()));
auto numSubDevices = deviceBitfield.count();
this->allocatePrivateMemoryPerDispatch = modulePrivateMemorySize * numSubDevices > globalMemorySize;
PRINT_DEBUG_STRING(NEO::debugManager.flags.PrintDebugMessages.get(), stderr, "Private Memory Per Dispatch %d for modulePrivateMemorySize %zu subDevices %zu globalMemorySize %" PRIu64 "\n",
this->allocatePrivateMemoryPerDispatch, modulePrivateMemorySize, numSubDevices, globalMemorySize);
}
}

View File

@@ -1,5 +1,5 @@
/*
* Copyright (C) 2022-2024 Intel Corporation
* Copyright (C) 2022-2025 Intel Corporation
*
* SPDX-License-Identifier: MIT
*
@@ -66,6 +66,10 @@ int main(int argc, char *argv[]) {
ze_kernel_desc_t kernelDesc = {ZE_STRUCTURE_TYPE_KERNEL_DESC};
kernelDesc.pKernelName = "memcpy_bytes";
SUCCESS_OR_TERMINATE(zeKernelCreate(module, &kernelDesc, &kernel));
ze_kernel_properties_t kernProps = {ZE_STRUCTURE_TYPE_KERNEL_PROPERTIES};
SUCCESS_OR_TERMINATE(zeKernelGetProperties(kernel, &kernProps));
std::cout << "function pointer memcpy_bytes Private size = " << std::dec << kernProps.privateMemSize << std::endl;
std::cout << "function pointer memcpy_bytes Spill size = " << std::dec << kernProps.spillMemSize << std::endl;
uint32_t groupSizeX = 1u;
uint32_t groupSizeY = 1u;