intel/compute-runtime
1
0
Fork 0
mirror of https://github.com/intel/compute-runtime.git synced 2025-12-19 06:24:51 +08:00
Code Issues Packages Projects Releases Wiki Activity

refactor: engineInstancedType removed

Browse Source 
Related-To: NEO-12594

Signed-off-by: Wojciech Konior <wojciech.konior@intel.com>
This commit is contained in:
Wojciech Konior
2024-10-08 13:23:41 +00:00
committed by Compute-Runtime-Automation
parent 834f00d48b
commit 6b40f9bc5a
65 changed files with 195 additions and 966 deletions
Download Patch File Download Diff File Expand all files Collapse all files

2
level_zero/api/core/ze_module_api_entrypoints.h
View File

@@ -103,7 +103,7 @@ ze_result_t zeKernelSuggestGroupSize(
ze_result_t zeKernelSuggestMaxCooperativeGroupCount(
ze_kernel_handle_t hKernel,
uint32_t *totalGroupCount) {
*totalGroupCount = L0::Kernel::fromHandle(hKernel)->suggestMaxCooperativeGroupCount(NEO::EngineGroupType::compute, false, false);
*totalGroupCount = L0::Kernel::fromHandle(hKernel)->suggestMaxCooperativeGroupCount(NEO::EngineGroupType::compute, false);
return ZE_RESULT_SUCCESS;
}

2
level_zero/core/source/cmdlist/cmdlist_hw.inl
View File

@@ -2771,7 +2771,7 @@ void CommandListCoreFamily<gfxCoreFamily>::appendSignalInOrderDependencyCounter(
template <GFXCORE_FAMILY gfxCoreFamily>
ze_result_t CommandListCoreFamily<gfxCoreFamily>::programSyncBuffer(Kernel &kernel, NEO::Device &device,
const ze_group_count_t &threadGroupDimensions) {
uint32_t maximalNumberOfWorkgroupsAllowed = kernel.suggestMaxCooperativeGroupCount(this->engineGroupType, device.isEngineInstanced(), false);
uint32_t maximalNumberOfWorkgroupsAllowed = kernel.suggestMaxCooperativeGroupCount(this->engineGroupType, false);
size_t requestedNumberOfWorkgroups = (threadGroupDimensions.groupCountX * threadGroupDimensions.groupCountY * threadGroupDimensions.groupCountZ);
if (requestedNumberOfWorkgroups > maximalNumberOfWorkgroupsAllowed) {

2
level_zero/core/source/cmdlist/cmdlist_imp.cpp
View File

@@ -292,7 +292,7 @@ void CommandListImp::setStreamPropertiesDefaultSettings(NEO::StreamProperties &s
streamProperties.stateComputeMode.setPropertiesCoherencyDevicePreemption(cmdListDefaultCoherency, this->commandListPreemptionMode, true);
}
streamProperties.frontEndState.setPropertiesDisableOverdispatchEngineInstanced(cmdListDefaultDisableOverdispatch, cmdListDefaultEngineInstancedDevice, true);
streamProperties.frontEndState.setPropertiesDisableOverdispatch(cmdListDefaultDisableOverdispatch, true);
streamProperties.pipelineSelect.setPropertiesModeSelectedMediaSamplerClockGate(cmdListDefaultPipelineSelectModeSelected, cmdListDefaultMediaSamplerClockGate, true);
}

1
level_zero/core/source/cmdlist/cmdlist_imp.h
View File

@@ -56,7 +56,6 @@ struct CommandListImp : public CommandList {
~CommandListImp() override = default;
static constexpr int32_t cmdListDefaultEngineInstancedDevice = NEO::StreamProperty::initValue;
static constexpr bool cmdListDefaultCoherency = false;
static constexpr bool cmdListDefaultDisableOverdispatch = true;
static constexpr bool cmdListDefaultPipelineSelectModeSelected = true;

3
level_zero/core/source/cmdqueue/cmdqueue_hw.h
View File

@@ -33,7 +33,7 @@ struct CommandQueueHw : public CommandQueueImp {
size_t estimateStateBaseAddressCmdSize();
MOCKABLE_VIRTUAL void programFrontEnd(uint64_t scratchAddress, uint32_t perThreadScratchSpaceSlot0Size, NEO::LinearStream &commandStream, NEO::StreamProperties &streamProperties);
MOCKABLE_VIRTUAL size_t estimateFrontEndCmdSizeForMultipleCommandLists(bool &isFrontEndStateDirty, int32_t engineInstanced, CommandList *commandList,
MOCKABLE_VIRTUAL size_t estimateFrontEndCmdSizeForMultipleCommandLists(bool &isFrontEndStateDirty, CommandList *commandList,
NEO::StreamProperties &csrState,
const NEO::StreamProperties &cmdListRequired,
const NEO::StreamProperties &cmdListFinal,
@@ -88,7 +88,6 @@ struct CommandQueueHw : public CommandQueueImp {
NEO::PreemptionMode statePreemption{};
uint32_t perThreadScratchSpaceSlot0Size = 0;
uint32_t perThreadScratchSpaceSlot1Size = 0;
int32_t engineInstanced = -1;
UnifiedMemoryControls unifiedMemoryControls{};
bool anyCommandListWithCooperativeKernels = false;

11
level_zero/core/source/cmdqueue/cmdqueue_hw.inl
View File

@@ -617,7 +617,7 @@ size_t CommandQueueHw<gfxCoreFamily>::estimateFrontEndCmdSize(bool isFrontEndDir
template <GFXCORE_FAMILY gfxCoreFamily>
size_t CommandQueueHw<gfxCoreFamily>::estimateFrontEndCmdSizeForMultipleCommandLists(
bool &isFrontEndStateDirty, int32_t engineInstanced, CommandList *commandList,
bool &isFrontEndStateDirty, CommandList *commandList,
NEO::StreamProperties &csrState,
const NEO::StreamProperties &cmdListRequired,
const NEO::StreamProperties &cmdListFinal,
@@ -635,7 +635,7 @@ size_t CommandQueueHw<gfxCoreFamily>::estimateFrontEndCmdSizeForMultipleCommandL
if (isFrontEndStateDirty) {
csrState.frontEndState.copyPropertiesAll(cmdListRequired.frontEndState);
csrState.frontEndState.setPropertySingleSliceDispatchCcsMode(engineInstanced);
csrState.frontEndState.setPropertySingleSliceDispatchCcsMode();
} else {
csrState.frontEndState.copyPropertiesComputeDispatchAllWalkerEnableDisableEuFusion(cmdListRequired.frontEndState);
feCurrentDirty = csrState.frontEndState.isDirty();
@@ -906,14 +906,11 @@ size_t CommandQueueHw<gfxCoreFamily>::estimateCommandListResidencySize(CommandLi
template <GFXCORE_FAMILY gfxCoreFamily>
void CommandQueueHw<gfxCoreFamily>::setFrontEndStateProperties(CommandListExecutionContext &ctx) {
auto isEngineInstanced = csr->getOsContext().isEngineInstanced();
auto &streamProperties = this->csr->getStreamProperties();
if (!frontEndTrackingEnabled()) {
streamProperties.frontEndState.setPropertiesAll(ctx.anyCommandListWithCooperativeKernels, ctx.anyCommandListRequiresDisabledEUFusion,
true, isEngineInstanced);
true);
ctx.frontEndStateDirty |= streamProperties.frontEndState.isDirty();
} else {
ctx.engineInstanced = isEngineInstanced;
}
ctx.frontEndStateDirty |= csr->getMediaVFEStateDirty();
ctx.globalInit |= ctx.frontEndStateDirty;
@@ -971,7 +968,7 @@ size_t CommandQueueHw<gfxCoreFamily>::estimateLinearStreamSizeComplementary(
const NEO::StreamProperties &requiredStreamState = cmdList->getRequiredStreamState();
const NEO::StreamProperties &finalStreamState = cmdList->getFinalStreamState();
linearStreamSizeEstimate += estimateFrontEndCmdSizeForMultipleCommandLists(frontEndStateDirty, ctx.engineInstanced, cmdList,
linearStreamSizeEstimate += estimateFrontEndCmdSizeForMultipleCommandLists(frontEndStateDirty, cmdList,
streamProperties, requiredStreamState, finalStreamState,
cmdListState.requiredState,
cmdListState.flags.propertyFeDirty, cmdListState.flags.frontEndReturnPoint);

2
level_zero/core/source/cmdqueue/cmdqueue_xe_hp_core_and_later.inl
View File

@@ -181,7 +181,7 @@ void CommandQueueHw<gfxCoreFamily>::patchCommands(CommandList &commandList, uint
cfeStateCmd = reinterpret_cast<CFE_STATE *>(commandToPatch.pCommand);
cfeStateCmd->setScratchSpaceBuffer(lowScratchAddress);
NEO::PreambleHelper<GfxFamily>::setSingleSliceDispatchMode(cfeStateCmd, csr->getOsContext().isEngineInstanced());
NEO::PreambleHelper<GfxFamily>::setSingleSliceDispatchMode(cfeStateCmd, false);
*reinterpret_cast<CFE_STATE *>(commandToPatch.pDestination) = *cfeStateCmd;
break;

2
level_zero/core/source/driver/driver_handle_imp.cpp
View File

@@ -349,7 +349,7 @@ ze_result_t DriverHandleImp::parseAffinityMaskCombined(uint32_t *pCount, ze_devi
continue;
}
// Combined Device Hierarchy not supported with AllowSingleTileEngineInstancedSubDevices
// Combined Device Hierarchy
// so ignore X.Y
if (subEntries.size() > 1) {
continue;

5
level_zero/core/source/fabric/fabric.cpp
View File

@@ -25,11 +25,6 @@ FabricVertex::~FabricVertex() {
FabricVertex *FabricVertex::createFromDevice(Device *device) {
// Fabric Vertices are not created for engine instanced devices
if (device->getNEODevice()->isEngineInstanced()) {
return nullptr;
}
auto fabricVertex = new FabricVertex();
UNRECOVERABLE_IF(fabricVertex == nullptr);

2
level_zero/core/source/kernel/kernel.h
View File

@@ -145,7 +145,7 @@ struct Kernel : _ze_kernel_handle_t, virtual NEO::DispatchKernelEncoderI {
virtual void patchGlobalOffset() = 0;
virtual void patchRegionParams(const CmdListKernelLaunchParams &launchParams) = 0;
virtual uint32_t suggestMaxCooperativeGroupCount(NEO::EngineGroupType engineGroupType, bool isEngineInstanced, bool forceSingleTileQuery) = 0;
virtual uint32_t suggestMaxCooperativeGroupCount(NEO::EngineGroupType engineGroupType, bool forceSingleTileQuery) = 0;
virtual ze_result_t setCacheConfig(ze_cache_config_flags_t flags) = 0;
virtual ze_result_t getProfileInfo(zet_profile_properties_t *pProfileProperties) = 0;

12
level_zero/core/source/kernel/kernel_imp.cpp
View File

@@ -431,13 +431,12 @@ ze_result_t KernelImp::setGroupSize(uint32_t groupSizeX, uint32_t groupSizeY,
}
if (this->heaplessEnabled && this->localDispatchSupport) {
auto isEngineIstanced = neoDevice->isEngineInstanced();
this->maxWgCountPerTileCcs = suggestMaxCooperativeGroupCount(NEO::EngineGroupType::compute, isEngineIstanced, true);
this->maxWgCountPerTileCcs = suggestMaxCooperativeGroupCount(NEO::EngineGroupType::compute, true);
if (this->rcsAvailable) {
this->maxWgCountPerTileRcs = suggestMaxCooperativeGroupCount(NEO::EngineGroupType::renderCompute, isEngineIstanced, true);
this->maxWgCountPerTileRcs = suggestMaxCooperativeGroupCount(NEO::EngineGroupType::renderCompute, true);
}
if (this->cooperativeSupport) {
this->maxWgCountPerTileCooperative = suggestMaxCooperativeGroupCount(NEO::EngineGroupType::cooperativeCompute, isEngineIstanced, true);
this->maxWgCountPerTileCooperative = suggestMaxCooperativeGroupCount(NEO::EngineGroupType::cooperativeCompute, true);
}
}
return ZE_RESULT_SUCCESS;
@@ -503,7 +502,7 @@ ze_result_t KernelImp::suggestGroupSize(uint32_t globalSizeX, uint32_t globalSiz
return ZE_RESULT_SUCCESS;
}
uint32_t KernelImp::suggestMaxCooperativeGroupCount(NEO::EngineGroupType engineGroupType, uint32_t *groupSize, bool isEngineInstanced, bool forceSingleTileQuery) {
uint32_t KernelImp::suggestMaxCooperativeGroupCount(NEO::EngineGroupType engineGroupType, uint32_t *groupSize, bool forceSingleTileQuery) {
auto &rootDeviceEnvironment = module->getDevice()->getNEODevice()->getRootDeviceEnvironment();
auto &helper = rootDeviceEnvironment.getHelper<NEO::GfxCoreHelper>();
auto &descriptor = kernelImmData->getDescriptor();
@@ -525,8 +524,7 @@ uint32_t KernelImp::suggestMaxCooperativeGroupCount(NEO::EngineGroupType engineG
usedSlmSize,
workDim,
localWorkSize,
engineGroupType,
isEngineInstanced);
engineGroupType);
}
ze_result_t KernelImp::setIndirectAccess(ze_kernel_indirect_access_flags_t flags) {

6
level_zero/core/source/kernel/kernel_imp.h
View File

@@ -70,15 +70,15 @@ struct KernelImp : Kernel {
ze_result_t getKernelName(size_t *pSize, char *pName) override;
uint32_t suggestMaxCooperativeGroupCount(NEO::EngineGroupType engineGroupType, bool isEngineInstanced, bool forceSingleTileQuery) override {
uint32_t suggestMaxCooperativeGroupCount(NEO::EngineGroupType engineGroupType, bool forceSingleTileQuery) override {
UNRECOVERABLE_IF(0 == this->groupSize[0]);
UNRECOVERABLE_IF(0 == this->groupSize[1]);
UNRECOVERABLE_IF(0 == this->groupSize[2]);
return suggestMaxCooperativeGroupCount(engineGroupType, this->groupSize, isEngineInstanced, forceSingleTileQuery);
return suggestMaxCooperativeGroupCount(engineGroupType, this->groupSize, forceSingleTileQuery);
}
uint32_t suggestMaxCooperativeGroupCount(NEO::EngineGroupType engineGroupType, uint32_t *groupSize, bool isEngineInstanced, bool forceSingleTileQuery);
uint32_t suggestMaxCooperativeGroupCount(NEO::EngineGroupType engineGroupType, uint32_t *groupSize, bool forceSingleTileQuery);
const uint8_t *getCrossThreadData() const override { return crossThreadData.get(); }
uint32_t getCrossThreadDataSize() const override { return crossThreadDataSize; }

4
level_zero/core/test/unit_tests/fixtures/kernel_max_cooperative_groups_count_fixture.cpp
View File

@@ -44,8 +44,8 @@ uint32_t KernelImpSuggestMaxCooperativeGroupCountFixture::getMaxWorkGroupCount()
kernel.groupSize[0] = lws[0];
kernel.groupSize[1] = lws[1];
kernel.groupSize[2] = lws[2];
uint32_t totalGroupCount = kernel.KernelImp::suggestMaxCooperativeGroupCount(NEO::EngineGroupType::cooperativeCompute, true, false);
uint32_t totalGroupCount = kernel.KernelImp::suggestMaxCooperativeGroupCount(NEO::EngineGroupType::cooperativeCompute, false);
return totalGroupCount;
}
} // namespace ult
} // namespace L0
} // namespace L0

2
level_zero/core/test/unit_tests/sources/cmdlist/test_cmdlist_7.cpp
View File

@@ -2144,7 +2144,7 @@ TEST(CommandList, givenContextGroupEnabledWhenCreatingImmediateCommandListWithIn
auto device = driverHandle->devices[0];
for (auto &engine : neoDevice->secondaryEngines[aub_stream::ENGINE_CCS].engines) {
EngineDescriptor descriptor({aub_stream::ENGINE_CCS, engine.osContext->getEngineUsage()}, engine.osContext->getDeviceBitfield(), PreemptionMode::Disabled, false, false);
EngineDescriptor descriptor({aub_stream::ENGINE_CCS, engine.osContext->getEngineUsage()}, engine.osContext->getDeviceBitfield(), PreemptionMode::Disabled, false);
auto newOsContext = new MockOsContext(0, 0, descriptor);
mockOsContexts.push_back(newOsContext);
newOsContext->incRefInternal();

2
level_zero/core/test/unit_tests/sources/cmdlist/test_cmdlist_append_launch_kernel_3.cpp
View File

@@ -498,7 +498,7 @@ HWTEST2_F(CommandListAppendLaunchKernel, givenKernelUsingSyncBufferWhenAppendLau
}
{
VariableBackup<uint32_t> groupCountX{&groupCount.groupCountX};
uint32_t maximalNumberOfWorkgroupsAllowed = kernel.suggestMaxCooperativeGroupCount(engineGroupType, false, false);
uint32_t maximalNumberOfWorkgroupsAllowed = kernel.suggestMaxCooperativeGroupCount(engineGroupType, false);
groupCountX = maximalNumberOfWorkgroupsAllowed + 1;
pCommandList = std::make_unique<WhiteBox<::L0::CommandListCoreFamily<gfxCoreFamily>>>();
pCommandList->initialize(device, engineGroupType, 0u);

4
level_zero/core/test/unit_tests/sources/cmdqueue/test_cmdqueue_1.cpp
View File

@@ -796,12 +796,12 @@ TEST_F(DeviceCreateCommandQueueTest, givenLowPriorityDescWhenCreateCommandQueueI
TEST_F(DeviceCreateCommandQueueTest, givenCopyOrdinalWhenCreateCommandQueueWithLowPriorityDescIsCalledThenCopyCsrIsAssigned) {
auto copyCsr = std::unique_ptr<NEO::CommandStreamReceiver>(neoDevice->createCommandStreamReceiver());
EngineDescriptor copyEngineDescriptor({aub_stream::ENGINE_BCS, EngineUsage::regular}, neoDevice->getDeviceBitfield(), neoDevice->getPreemptionMode(), false, false);
EngineDescriptor copyEngineDescriptor({aub_stream::ENGINE_BCS, EngineUsage::regular}, neoDevice->getDeviceBitfield(), neoDevice->getPreemptionMode(), false);
auto copyOsContext = neoDevice->getExecutionEnvironment()->memoryManager->createAndRegisterOsContext(copyCsr.get(), copyEngineDescriptor);
copyCsr->setupContext(*copyOsContext);
auto computeCsr = std::unique_ptr<NEO::CommandStreamReceiver>(neoDevice->createCommandStreamReceiver());
EngineDescriptor computeEngineDescriptor({aub_stream::ENGINE_CCS, EngineUsage::lowPriority}, neoDevice->getDeviceBitfield(), neoDevice->getPreemptionMode(), false, false);
EngineDescriptor computeEngineDescriptor({aub_stream::ENGINE_CCS, EngineUsage::lowPriority}, neoDevice->getDeviceBitfield(), neoDevice->getPreemptionMode(), false);
auto computeOsContext = neoDevice->getExecutionEnvironment()->memoryManager->createAndRegisterOsContext(computeCsr.get(), computeEngineDescriptor);
computeCsr->setupContext(*computeOsContext);

155
level_zero/core/test/unit_tests/sources/cmdqueue/test_cmdqueue_3.cpp
View File

@@ -630,57 +630,6 @@ HWTEST_F(CommandQueueIndirectAllocations, givenImmediateCommandListAndFlushTaskW
device->getDriverHandle()->getSvmAllocsManager()->freeSVMAlloc(deviceAlloc);
}
struct EngineInstancedDeviceExecuteTests : public ::testing::Test {
void SetUp() override {
debugManager.flags.EngineInstancedSubDevices.set(true);
}
bool createDevices(uint32_t numGenericSubDevices, uint32_t numCcs) {
debugManager.flags.CreateMultipleSubDevices.set(numGenericSubDevices);
auto executionEnvironment = std::make_unique<NEO::MockExecutionEnvironment>();
auto &rootDeviceEnvironment = *executionEnvironment->rootDeviceEnvironments[0].get();
auto hwInfo = rootDeviceEnvironment.getMutableHardwareInfo();
hwInfo->gtSystemInfo.CCSInfo.NumberOfCCSEnabled = numCcs;
hwInfo->featureTable.flags.ftrCCSNode = (numCcs > 0);
auto &gfxCoreHelper = rootDeviceEnvironment.getHelper<NEO::GfxCoreHelper>();
auto &productHelper = rootDeviceEnvironment.getHelper<NEO::ProductHelper>();
gfxCoreHelper.adjustDefaultEngineType(hwInfo, productHelper, rootDeviceEnvironment.ailConfiguration.get());
if (!multiCcsDevice(rootDeviceEnvironment, numCcs)) {
return false;
}
executionEnvironment->parseAffinityMask();
deviceFactory = std::make_unique<NEO::UltDeviceFactory>(1, numGenericSubDevices, *executionEnvironment.release());
rootDevice = deviceFactory->rootDevices[0];
EXPECT_NE(nullptr, rootDevice);
return true;
}
bool multiCcsDevice(const NEO::RootDeviceEnvironment &rootDeviceEnvironment, uint32_t expectedNumCcs) {
auto &gfxCoreHelper = rootDeviceEnvironment.getHelper<NEO::GfxCoreHelper>();
auto gpgpuEngines = gfxCoreHelper.getGpgpuEngineInstances(rootDeviceEnvironment);
uint32_t numCcs = 0;
for (auto &engine : gpgpuEngines) {
if (EngineHelpers::isCcs(engine.first) && (engine.second == EngineUsage::regular)) {
numCcs++;
}
}
return (numCcs == expectedNumCcs);
}
DebugManagerStateRestore restorer;
std::unique_ptr<NEO::UltDeviceFactory> deviceFactory;
NEO::MockDevice *rootDevice = nullptr;
};
struct SingleSliceDispatchSupportMatcher {
template <PRODUCT_FAMILY productFamily>
static constexpr bool isMatched() {
@@ -689,106 +638,6 @@ struct SingleSliceDispatchSupportMatcher {
}
};
HWTEST2_F(EngineInstancedDeviceExecuteTests, givenEngineInstancedDeviceWhenExecutingThenEnableSingleSliceDispatch, SingleSliceDispatchSupportMatcher) {
using CFE_STATE = typename FamilyType::CFE_STATE;
constexpr uint32_t genericDevicesCount = 1;
constexpr uint32_t ccsCount = 2;
debugManager.flags.AllowSingleTileEngineInstancedSubDevices.set(true);
if (!createDevices(genericDevicesCount, ccsCount)) {
GTEST_SKIP();
}
auto subDevice = static_cast<MockSubDevice *>(rootDevice->getSubDevice(0));
auto defaultEngine = subDevice->getDefaultEngine();
EXPECT_TRUE(defaultEngine.osContext->isEngineInstanced());
std::vector<std::unique_ptr<NEO::Device>> devices;
devices.push_back(std::unique_ptr<NEO::Device>(subDevice));
auto driverHandle = std::make_unique<Mock<L0::DriverHandleImp>>();
driverHandle->initialize(std::move(devices));
auto l0Device = driverHandle->devices[0];
ze_command_queue_desc_t desc = {};
NEO::CommandStreamReceiver *csr;
l0Device->getCsrForOrdinalAndIndex(&csr, 0u, 0u, ZE_COMMAND_QUEUE_PRIORITY_NORMAL, false);
ze_result_t returnValue;
auto commandQueue = whiteboxCast(CommandQueue::create(productFamily, l0Device, csr, &desc, false, false, false, returnValue));
auto commandList = std::unique_ptr<CommandList>(CommandList::whiteboxCast(CommandList::create(productFamily, l0Device, NEO::EngineGroupType::compute, 0u, returnValue, false)));
auto commandListHandle = commandList->toHandle();
commandList->close();
commandQueue->executeCommandLists(1, &commandListHandle, nullptr, false, nullptr);
GenCmdList cmdList;
FamilyType::Parse::parseCommandBuffer(cmdList, commandQueue->commandStream.getCpuBase(), commandQueue->commandStream.getUsed());
auto cfeStates = findAll<CFE_STATE *>(cmdList.begin(), cmdList.end());
EXPECT_NE(0u, cfeStates.size());
for (auto &cmd : cfeStates) {
auto cfeState = reinterpret_cast<CFE_STATE *>(*cmd);
EXPECT_TRUE(cfeState->getSingleSliceDispatchCcsMode());
}
commandQueue->destroy();
}
HWTEST2_F(EngineInstancedDeviceExecuteTests, givenEngineInstancedDeviceWithFabricEnumerationWhenExecutingThenEnableSingleSliceDispatch, SingleSliceDispatchSupportMatcher) {
using CFE_STATE = typename FamilyType::CFE_STATE;
constexpr uint32_t genericDevicesCount = 1;
constexpr uint32_t ccsCount = 2;
debugManager.flags.AllowSingleTileEngineInstancedSubDevices.set(true);
if (!createDevices(genericDevicesCount, ccsCount)) {
GTEST_SKIP();
}
auto subDevice = static_cast<MockSubDevice *>(rootDevice->getSubDevice(0));
auto defaultEngine = subDevice->getDefaultEngine();
EXPECT_TRUE(defaultEngine.osContext->isEngineInstanced());
std::vector<std::unique_ptr<NEO::Device>> devices;
devices.push_back(std::unique_ptr<NEO::Device>(subDevice));
auto driverHandle = std::make_unique<Mock<L0::DriverHandleImp>>();
driverHandle->initialize(std::move(devices));
driverHandle->initializeVertexes();
auto l0Device = driverHandle->devices[0];
ze_command_queue_desc_t desc = {};
NEO::CommandStreamReceiver *csr;
l0Device->getCsrForOrdinalAndIndex(&csr, 0u, 0u, ZE_COMMAND_QUEUE_PRIORITY_NORMAL, false);
ze_result_t returnValue;
auto commandQueue = whiteboxCast(CommandQueue::create(productFamily, l0Device, csr, &desc, false, false, false, returnValue));
auto commandList = std::unique_ptr<CommandList>(CommandList::whiteboxCast(CommandList::create(productFamily, l0Device, NEO::EngineGroupType::compute, 0u, returnValue, false)));
auto commandListHandle = commandList->toHandle();
commandList->close();
commandQueue->executeCommandLists(1, &commandListHandle, nullptr, false, nullptr);
GenCmdList cmdList;
FamilyType::Parse::parseCommandBuffer(cmdList, commandQueue->commandStream.getCpuBase(), commandQueue->commandStream.getUsed());
auto cfeStates = findAll<CFE_STATE *>(cmdList.begin(), cmdList.end());
EXPECT_NE(0u, cfeStates.size());
for (auto &cmd : cfeStates) {
auto cfeState = reinterpret_cast<CFE_STATE *>(*cmd);
EXPECT_TRUE(cfeState->getSingleSliceDispatchCcsMode());
}
commandQueue->destroy();
}
HWTEST2_F(CommandQueueIndirectAllocations, givenCtxWithIndirectAccessWhenExecutingCommandListImmediateWithFlushTaskThenHandleIndirectAccessCalled, MatchAny) {
ze_command_queue_desc_t desc = {};
auto csr = neoDevice->getDefaultEngine().commandStreamReceiver;
@@ -1050,7 +899,7 @@ size_t estimateAllCommmandLists(MockCommandQueueHw<gfxCoreFamily> *commandQueue,
auto cmdListPtr = CommandList::fromHandle(commandListHandles[i]);
auto requiredState = cmdListPtr->getRequiredStreamState();
auto finalState = cmdListPtr->getFinalStreamState();
estimatedSize += commandQueue->estimateFrontEndCmdSizeForMultipleCommandLists(frontEndStateDirty, -1, cmdListPtr, csrStateCopy, requiredState, finalState, dummyProperties, feDirty, feRetPoint);
estimatedSize += commandQueue->estimateFrontEndCmdSizeForMultipleCommandLists(frontEndStateDirty, cmdListPtr, csrStateCopy, requiredState, finalState, dummyProperties, feDirty, feRetPoint);
}
return estimatedSize;
}
@@ -1481,7 +1330,7 @@ TEST(CommandQueue, givenContextGroupEnabledWhenCreatingCommandQueuesWithInterrup
auto device = driverHandle->devices[0];
for (auto &engine : neoDevice->secondaryEngines[aub_stream::ENGINE_CCS].engines) {
EngineDescriptor descriptor({aub_stream::ENGINE_CCS, engine.osContext->getEngineUsage()}, engine.osContext->getDeviceBitfield(), PreemptionMode::Disabled, false, false);
EngineDescriptor descriptor({aub_stream::ENGINE_CCS, engine.osContext->getEngineUsage()}, engine.osContext->getDeviceBitfield(), PreemptionMode::Disabled, false);
auto newOsContext = new MockOsContext(0, 0, descriptor);
mockOsContexts.push_back(newOsContext);
newOsContext->incRefInternal();

6
level_zero/core/test/unit_tests/sources/fabric/test_fabric.cpp
View File

@@ -156,17 +156,13 @@ TEST_F(FabricVertexFixture, givenFabricVerticesAreCreatedWhenzeFabricVertexGetPr
}
}
TEST(FabricEngineInstanceTest, GivenEngineInstancedDeviceWhenFabricVerticesAreCreatedThenSkipCreationForEngineInstanced) {
TEST(FabricEngineInstanceTest, GivenSubDeviceWhenFabricVerticesAreCreatedThenSkipCreationForSubDevice) {
auto hwInfo = *NEO::defaultHwInfo.get();
auto executionEnvironment = NEO::MockDevice::prepareExecutionEnvironment(&hwInfo, 0u);
std::vector<std::unique_ptr<NEO::Device>> devices(1);
devices[0].reset(static_cast<NEO::Device *>(NEO::MockDevice::createWithExecutionEnvironment<NEO::MockDevice>(&hwInfo, executionEnvironment, 0u)));
auto mockDevice = static_cast<NEO::MockDevice *>(devices[0].get());
NEO::SubDevice *subDevice = static_cast<NEO::SubDevice *>(mockDevice->createEngineInstancedSubDevice(0, defaultHwInfo->capabilityTable.defaultEngineType));
mockDevice->subdevices.push_back(subDevice);
std::unique_ptr<Mock<L0::DriverHandleImp>> driverHandle;
driverHandle = std::make_unique<Mock<L0::DriverHandleImp>>();
ze_result_t res = driverHandle->initialize(std::move(devices));

11
opencl/source/cl_device/cl_device.cpp
View File

@@ -1,5 +1,5 @@
/*
* Copyright (C) 2020-2023 Intel Corporation
* Copyright (C) 2020-2024 Intel Corporation
*
* SPDX-License-Identifier: MIT
*
@@ -117,15 +117,6 @@ ClDevice *ClDevice::getSubDevice(uint32_t deviceId) const {
}
ClDevice *ClDevice::getNearestGenericSubDevice(uint32_t deviceId) {
/*
* EngineInstanced: Upper level
* Generic SubDevice: 'this'
* RootCsr Device: Next level SubDevice (generic)
*/
if (getDevice().isEngineInstanced()) {
return rootClDevice.getNearestGenericSubDevice(Math::log2(static_cast<uint32_t>(getDeviceBitfield().to_ulong())));
}
if (subDevices.empty() || !getDevice().hasRootCsr()) {
return this;

3
opencl/source/cl_device/cl_device_caps.cpp
View File

@@ -296,9 +296,6 @@ void ClDevice::initializeCaps() {
deviceInfo.maxWorkItemDimensions = 3;
deviceInfo.maxComputUnits = systemInfo.EUCount * subDevicesCount;
if (device.isEngineInstanced()) {
deviceInfo.maxComputUnits /= systemInfo.CCSInfo.NumberOfCCSEnabled;
}
deviceInfo.maxConstantArgs = 8;
deviceInfo.maxSliceCount = systemInfo.SliceCount;

4
opencl/source/kernel/kernel.cpp
View File

@@ -1182,7 +1182,6 @@ uint32_t Kernel::getMaxWorkGroupCount(const cl_uint workDim, const size_t *local
auto engineGroupType = helper.getEngineGroupType(commandQueue->getGpgpuEngine().getEngineType(),
commandQueue->getGpgpuEngine().getEngineUsage(), hardwareInfo);
auto isEngineInstanced = commandQueue->getGpgpuCommandStreamReceiver().getOsContext().isEngineInstanced();
auto usedSlmSize = helper.alignSlmSize(slmTotalSize);
@@ -1201,8 +1200,7 @@ uint32_t Kernel::getMaxWorkGroupCount(const cl_uint workDim, const size_t *local
usedSlmSize,
workDim,
localWorkSize,
engineGroupType,
isEngineInstanced);
engineGroupType);
return maxWorkGroupCount;
}

3
opencl/test/unit_test/api/cl_create_sub_devices_tests.inl
View File

@@ -1,5 +1,5 @@
/*
* Copyright (C) 2018-2023 Intel Corporation
* Copyright (C) 2018-2024 Intel Corporation
*
* SPDX-License-Identifier: MIT
*
@@ -372,7 +372,6 @@ TEST_F(ClCreateSubDevicesDeviceInfoTests, WhenGettingSubDeviceRelatedDeviceInfoV
}
TEST_F(ClCreateSubDevicesDeviceInfoTests, GivenRootDeviceWithoutSubDevicesWhenGettingSubDeviceRelatedDeviceInfoViaApiThenCorrectValuesAreSet) {
debugManager.flags.EngineInstancedSubDevices.set(false);
setup(1);
clGetDeviceInfo(device.get(), CL_DEVICE_PARENT_DEVICE,

6
opencl/test/unit_test/api/cl_enqueue_verify_memory.inl
View File

@@ -1,5 +1,5 @@
/*
* Copyright (C) 2019-2023 Intel Corporation
* Copyright (C) 2019-2024 Intel Corporation
*
* SPDX-License-Identifier: MIT
*
@@ -70,8 +70,8 @@ HWTEST_F(ClEnqueueVerifyMemoryIntelTests, givenActiveBcsEngineWhenCallingExpectM
UltCommandStreamReceiver<FamilyType> ultCsrBcs0(*pDevice->getExecutionEnvironment(), pDevice->getRootDeviceIndex(), pDevice->getDeviceBitfield());
UltCommandStreamReceiver<FamilyType> ultCsrBcs1(*pDevice->getExecutionEnvironment(), pDevice->getRootDeviceIndex(), pDevice->getDeviceBitfield());
MockOsContext osContext0(0, {{aub_stream::ENGINE_BCS, EngineUsage::regular}, 0, PreemptionMode::Disabled, false, false});
MockOsContext osContext1(1, {{aub_stream::ENGINE_BCS1, EngineUsage::regular}, 0, PreemptionMode::Disabled, false, false});
MockOsContext osContext0(0, {{aub_stream::ENGINE_BCS, EngineUsage::regular}, 0, PreemptionMode::Disabled, false});
MockOsContext osContext1(1, {{aub_stream::ENGINE_BCS1, EngineUsage::regular}, 0, PreemptionMode::Disabled, false});
EngineControl engineControl[2] = {{&ultCsrBcs0, &osContext0}, {&ultCsrBcs1, &osContext1}};

20
opencl/test/unit_test/device/device_tests.cpp
View File

@@ -307,12 +307,8 @@ TEST(DeviceCreation, givenMultiRootDeviceWhenTheyAreCreatedThenEachOsContextHasU
const auto &numGpgpuEngines = static_cast<uint32_t>(gfxCoreHelper.getGpgpuEngineInstances(device1->getRootDeviceEnvironment()).size());
size_t numExpectedGenericEnginesPerDevice = numGpgpuEngines;
size_t numExpectedEngineInstancedEnginesPerDevice = 0;
if (device1->getNumSubDevices() > 0) {
numExpectedEngineInstancedEnginesPerDevice = device1->getNumSubDevices();
}
auto expectedTotalRegisteredEnginesPerRootDevice = numExpectedGenericEnginesPerDevice + numExpectedEngineInstancedEnginesPerDevice;
auto expectedTotalRegisteredEnginesPerRootDevice = numExpectedGenericEnginesPerDevice;
uint32_t contextId = 0;
for (uint32_t i = 0; i < numDevices; i++) {
@@ -322,19 +318,6 @@ TEST(DeviceCreation, givenMultiRootDeviceWhenTheyAreCreatedThenEachOsContextHasU
EXPECT_EQ(expectedTotalRegisteredEnginesPerRootDevice, registeredEngines.size());
auto device = devices[i];
for (uint32_t j = 0; j < numExpectedEngineInstancedEnginesPerDevice; j++) {
auto subDevice = device->getSubDevice(j);
auto &engine = subDevice->getEngine(0);
EXPECT_EQ(contextId, engine.osContext->getContextId());
EXPECT_EQ(1u, engine.osContext->getDeviceBitfield().to_ulong());
EXPECT_EQ(registeredEngines[contextWithinRootDevice].commandStreamReceiver,
engine.commandStreamReceiver);
contextId++;
contextWithinRootDevice++;
}
for (uint32_t j = 0; j < numExpectedGenericEnginesPerDevice; j++) {
auto &engine = device->getEngine(j);
@@ -772,7 +755,6 @@ TEST_P(MultipleDeviceTest, givenMultipleDevicesWhenGetNumTilesThenReturnNumberOf
ultHwConfig.useMockedPrepareDeviceEnvironmentsFunc = false;
auto numDevices = GetParam();
debugManager.flags.CreateMultipleSubDevices.set(numDevices);
debugManager.flags.EngineInstancedSubDevices.set(false);
debugManager.flags.DeferOsContextInitialization.set(0);
initPlatform();
auto device = platform()->getClDevice(0);

425
opencl/test/unit_test/device/sub_device_tests.cpp
View File

@@ -34,9 +34,6 @@ TEST(SubDevicesTest, givenDefaultConfigWhenCreateRootDeviceThenItDoesntContainSu
auto device = std::unique_ptr<MockDevice>(MockDevice::createWithNewExecutionEnvironment<MockDevice>(defaultHwInfo.get()));
EXPECT_EQ(0u, device->getNumGenericSubDevices());
if (device->getNumSubDevices() > 0) {
EXPECT_TRUE(device->getSubDevice(0)->isEngineInstanced());
}
}
TEST(SubDevicesTest, givenCreateMultipleSubDevicesFlagSetWhenCreateRootDeviceThenItsSubdevicesHaveProperRootIdSet) {
@@ -354,27 +351,7 @@ TEST(SubDevicesTest, givenRootDeviceWithSubDevicesWithoutLocalMemoryWhenGettingG
}
}
TEST(SubDevicesTest, whenCreatingEngineInstancedSubDeviceThenSetCorrectSubdeviceIndex) {
class MyRootDevice : public RootDevice {
public:
using RootDevice::createEngineInstancedSubDevice;
using RootDevice::RootDevice;
};
auto executionEnvironment = new MockExecutionEnvironment();
executionEnvironment->rootDeviceEnvironments[0]->initGmm();
DeviceFactory::createMemoryManagerFunc(*executionEnvironment);
auto rootDevice = std::unique_ptr<MyRootDevice>(Device::create<MyRootDevice>(executionEnvironment, 0));
auto subDevice = std::unique_ptr<SubDevice>(rootDevice->createEngineInstancedSubDevice(1, defaultHwInfo->capabilityTable.defaultEngineType));
ASSERT_NE(nullptr, subDevice.get());
EXPECT_EQ(2u, subDevice->getDeviceBitfield().to_ulong());
}
struct EngineInstancedDeviceTests : public ::testing::Test {
struct DeviceTests : public ::testing::Test {
bool createDevices(uint32_t numGenericSubDevices, uint32_t numCcs) {
debugManager.flags.CreateMultipleSubDevices.set(numGenericSubDevices);
@@ -406,19 +383,6 @@ struct EngineInstancedDeviceTests : public ::testing::Test {
device->allEngines[0].osContext->isRootDevice());
}
bool isEngineInstanced(MockSubDevice *subDevice, aub_stream::EngineType engineType, uint32_t subDeviceIndex, DeviceBitfield deviceBitfield) {
bool isEngineInstanced = !subDevice->allEngines[0].osContext->isRootDevice();
isEngineInstanced &= subDevice->engineInstanced;
isEngineInstanced &= (subDevice->getNumGenericSubDevices() == 0);
isEngineInstanced &= (subDevice->getNumSubDevices() == 0);
isEngineInstanced &= (engineType == subDevice->engineInstancedType);
isEngineInstanced &= (subDeviceIndex == subDevice->getSubDeviceIndex());
isEngineInstanced &= (deviceBitfield == subDevice->getDeviceBitfield());
isEngineInstanced &= (subDevice->getAllEngines().size() == 1);
return isEngineInstanced;
}
template <typename MockDeviceT>
bool hasAllEngines(MockDeviceT *device) {
auto &gfxCoreHelper = device->getGfxCoreHelper();
@@ -472,37 +436,7 @@ struct EngineInstancedDeviceTests : public ::testing::Test {
MockDevice *rootDevice = nullptr;
};
TEST_F(EngineInstancedDeviceTests, givenDebugFlagSetAndMoreThanOneCcsWhenCreatingRootDeviceWithoutGenericSubDevicesThenCreateEngineInstanced) {
constexpr uint32_t genericDevicesCount = 1;
constexpr uint32_t ccsCount = 2;
debugManager.flags.EngineInstancedSubDevices.set(true);
debugManager.flags.AllowSingleTileEngineInstancedSubDevices.set(true);
if (!createDevices(genericDevicesCount, ccsCount)) {
GTEST_SKIP();
}
auto &hwInfo = rootDevice->getHardwareInfo();
EXPECT_EQ(ccsCount, hwInfo.gtSystemInfo.CCSInfo.NumberOfCCSEnabled);
EXPECT_EQ(ccsCount, rootDevice->getNumSubDevices());
EXPECT_EQ(0u, rootDevice->getNumGenericSubDevices());
EXPECT_FALSE(hasRootCsrOnly(rootDevice));
EXPECT_TRUE(hasAllEngines(rootDevice));
for (uint32_t i = 0; i < ccsCount; i++) {
auto engineType = static_cast<aub_stream::EngineType>(aub_stream::EngineType::ENGINE_CCS + i);
auto subDevice = static_cast<MockSubDevice *>(rootDevice->getSubDevice(i));
ASSERT_NE(nullptr, subDevice);
EXPECT_TRUE(isEngineInstanced(subDevice, engineType, 0, 1));
EXPECT_TRUE(hasEngineInstancedEngines(subDevice, engineType));
}
}
TEST_F(EngineInstancedDeviceTests, givenDebugFlagNotSetAndMoreThanOneCcsWhenCreatingRootDeviceWithoutGenericSubDevicesThenDontCreateEngineInstanced) {
TEST_F(DeviceTests, givenDebugFlagNotSetAndMoreThanOneCcsWhenCreatingRootDeviceWithoutGenericSubDevicesThenDontCreateSubDevice) {
constexpr uint32_t genericDevicesCount = 1;
constexpr uint32_t ccsCount = 2;
@@ -520,7 +454,7 @@ TEST_F(EngineInstancedDeviceTests, givenDebugFlagNotSetAndMoreThanOneCcsWhenCrea
EXPECT_TRUE(hasAllEngines(rootDevice));
}
TEST_F(EngineInstancedDeviceTests, givenDebugFlagSetAndZeroCcsesWhenCreatingRootDeviceWithoutGenericSubDevicesThenCreateEngineInstanced) {
TEST_F(DeviceTests, givenDebugFlagSetAndZeroCcsesWhenCreatingRootDeviceWithoutGenericSubDevicesThenCreateSubDevice) {
constexpr uint32_t genericDevicesCount = 1;
constexpr uint32_t ccsCount = 0;
@@ -534,7 +468,7 @@ TEST_F(EngineInstancedDeviceTests, givenDebugFlagSetAndZeroCcsesWhenCreatingRoot
EXPECT_FALSE(rootDevice->getNearestGenericSubDevice(0)->isSubDevice());
}
TEST_F(EngineInstancedDeviceTests, givenDebugFlagSetAndSingleCcsWhenCreatingRootDeviceWithoutGenericSubDevicesThenDontCreateEngineInstanced) {
TEST_F(DeviceTests, givenDebugFlagSetAndSingleCcsWhenCreatingRootDeviceWithoutGenericSubDevicesThenDontCreateSubDevice) {
constexpr uint32_t genericDevicesCount = 1;
constexpr uint32_t ccsCount = 1;
@@ -550,7 +484,7 @@ TEST_F(EngineInstancedDeviceTests, givenDebugFlagSetAndSingleCcsWhenCreatingRoot
EXPECT_FALSE(rootDevice->getNearestGenericSubDevice(0)->isSubDevice());
}
TEST_F(EngineInstancedDeviceTests, givenDebugFlagSetWhenCreatingRootDeviceWithGenericSubDevicesAndZeroCcsesThenDontCreateEngineInstanced) {
TEST_F(DeviceTests, givenDebugFlagSetWhenCreatingRootDeviceWithGenericSubDevicesAndZeroCcsesThenDontCreateSubDevice) {
constexpr uint32_t genericDevicesCount = 2;
constexpr uint32_t ccsCount = 0;
@@ -563,16 +497,14 @@ TEST_F(EngineInstancedDeviceTests, givenDebugFlagSetWhenCreatingRootDeviceWithGe
ASSERT_NE(nullptr, subDevice);
EXPECT_FALSE(subDevice->allEngines[0].osContext->isRootDevice());
EXPECT_FALSE(subDevice->engineInstanced);
EXPECT_EQ(0u, subDevice->getNumGenericSubDevices());
EXPECT_EQ(0u, subDevice->getNumSubDevices());
EXPECT_EQ(aub_stream::EngineType::NUM_ENGINES, subDevice->engineInstancedType);
EXPECT_TRUE(hasAllEngines(subDevice));
}
}
TEST_F(EngineInstancedDeviceTests, givenDebugFlagSetWhenCreatingRootDeviceWithGenericSubDevicesAndSingleCcsThenDontCreateEngineInstanced) {
TEST_F(DeviceTests, givenDebugFlagSetWhenCreatingRootDeviceWithGenericSubDevicesAndSingleCcsThenDontCreateSubDevice) {
constexpr uint32_t genericDevicesCount = 2;
constexpr uint32_t ccsCount = 1;
@@ -587,154 +519,47 @@ TEST_F(EngineInstancedDeviceTests, givenDebugFlagSetWhenCreatingRootDeviceWithGe
ASSERT_NE(nullptr, subDevice);
EXPECT_FALSE(subDevice->allEngines[0].osContext->isRootDevice());
EXPECT_FALSE(subDevice->engineInstanced);
EXPECT_EQ(0u, subDevice->getNumGenericSubDevices());
EXPECT_EQ(0u, subDevice->getNumSubDevices());
EXPECT_EQ(aub_stream::EngineType::NUM_ENGINES, subDevice->engineInstancedType);
EXPECT_TRUE(hasAllEngines(subDevice));
}
}
TEST_F(EngineInstancedDeviceTests, givenDebugFlagSetWhenCreatingRootDeviceWithGenericSubDevicesThenCreateEngineInstanced) {
constexpr uint32_t genericDevicesCount = 2;
constexpr uint32_t ccsCount = 2;
debugManager.flags.EngineInstancedSubDevices.set(true);
if (!createDevices(genericDevicesCount, ccsCount)) {
GTEST_SKIP();
}
EXPECT_TRUE(hasRootCsrOnly(rootDevice));
for (uint32_t i = 0; i < genericDevicesCount; i++) {
auto subDevice = static_cast<MockSubDevice *>(rootDevice->getSubDevice(i));
ASSERT_NE(nullptr, subDevice);
EXPECT_FALSE(subDevice->allEngines[0].osContext->isRootDevice());
EXPECT_FALSE(subDevice->engineInstanced);
EXPECT_EQ(0u, subDevice->getNumGenericSubDevices());
EXPECT_EQ(ccsCount, subDevice->getNumSubDevices());
EXPECT_EQ(aub_stream::EngineType::NUM_ENGINES, subDevice->engineInstancedType);
EXPECT_TRUE(hasAllEngines(subDevice));
for (uint32_t j = 0; j < ccsCount; j++) {
auto engineType = static_cast<aub_stream::EngineType>(aub_stream::EngineType::ENGINE_CCS + j);
auto engineSubDevice = static_cast<MockSubDevice *>(subDevice->getSubDevice(j));
ASSERT_NE(nullptr, engineSubDevice);
EXPECT_TRUE(isEngineInstanced(engineSubDevice, engineType, subDevice->getSubDeviceIndex(), subDevice->getDeviceBitfield()));
EXPECT_TRUE(hasEngineInstancedEngines(engineSubDevice, engineType));
}
}
}
TEST_F(EngineInstancedDeviceTests, givenEngineInstancedSubDeviceWhenEngineCreationFailsThenReturnFalse) {
TEST_F(DeviceTests, givenSubDeviceWhenEngineCreationFailsThenReturnFalse) {
constexpr uint32_t genericDevicesCount = 2;
constexpr uint32_t ccsCount = 0;
EXPECT_TRUE(createDevices(genericDevicesCount, ccsCount));
auto subDevice = static_cast<MockSubDevice *>(rootDevice->getSubDevice(0));
auto &gfxCoreHelper = rootDevice->getGfxCoreHelper();
auto &gpgpuEngines = gfxCoreHelper.getGpgpuEngineInstances(rootDevice->getRootDeviceEnvironment());
subDevice->engineInstanced = true;
subDevice->failOnCreateEngine = true;
subDevice->engineInstancedType = gpgpuEngines[0].first;
EXPECT_FALSE(subDevice->createEngines());
}
TEST_F(EngineInstancedDeviceTests, givenMultipleSubDevicesWhenCallingGetSubDeviceThenReturnCorrectObject) {
TEST_F(DeviceTests, givenMultipleClSubDevicesWhenCallingGetSubDeviceThenReturnCorrectObject) {
constexpr uint32_t genericDevicesCount = 2;
constexpr uint32_t ccsCount = 2;
debugManager.flags.EngineInstancedSubDevices.set(true);
if (!createDevices(genericDevicesCount, ccsCount)) {
GTEST_SKIP();
}
auto subDevice0 = rootDevice->subdevices[0];
auto subDevice1 = rootDevice->subdevices[1];
auto subSubDevice00 = subDevice0->getSubDevice(0);
auto subSubDevice01 = subDevice0->getSubDevice(1);
auto subSubDevice10 = subDevice1->getSubDevice(0);
auto subSubDevice11 = subDevice1->getSubDevice(1);
{
EXPECT_EQ(rootDevice->getSubDevice(0), subDevice0);
EXPECT_EQ(rootDevice->getNearestGenericSubDevice(0), subDevice0);
EXPECT_EQ(rootDevice->getSubDevice(1), subDevice1);
EXPECT_EQ(rootDevice->getNearestGenericSubDevice(1), subDevice1);
}
{
EXPECT_EQ(subDevice0->getNearestGenericSubDevice(0), subDevice0);
EXPECT_EQ(subDevice0->getNearestGenericSubDevice(1), subDevice0);
EXPECT_EQ(subDevice1->getNearestGenericSubDevice(0), subDevice1);
EXPECT_EQ(subDevice1->getNearestGenericSubDevice(1), subDevice1);
}
{
EXPECT_NE(subDevice0, subSubDevice00);
EXPECT_NE(subDevice0, subSubDevice01);
EXPECT_NE(subDevice1, subSubDevice10);
EXPECT_NE(subDevice1, subSubDevice11);
}
{
EXPECT_EQ(subSubDevice00->getNearestGenericSubDevice(0), subDevice0);
EXPECT_EQ(subSubDevice01->getNearestGenericSubDevice(0), subDevice0);
EXPECT_EQ(subSubDevice10->getNearestGenericSubDevice(0), subDevice1);
EXPECT_EQ(subSubDevice11->getNearestGenericSubDevice(0), subDevice1);
}
{
EXPECT_ANY_THROW(subSubDevice00->getSubDevice(0));
EXPECT_ANY_THROW(subSubDevice01->getSubDevice(0));
EXPECT_ANY_THROW(subSubDevice10->getSubDevice(0));
EXPECT_ANY_THROW(subSubDevice11->getSubDevice(0));
}
}
TEST_F(EngineInstancedDeviceTests, givenMultipleClSubDevicesWhenCallingGetSubDeviceThenReturnCorrectObject) {
constexpr uint32_t genericDevicesCount = 2;
constexpr uint32_t ccsCount = 2;
debugManager.flags.EngineInstancedSubDevices.set(true);
if (!createDevices(genericDevicesCount, ccsCount)) {
GTEST_SKIP();
}
auto subDevice = rootDevice->subdevices[0];
auto subSubDevice = subDevice->getSubDevice(0);
auto clRootDevice = std::make_unique<ClDevice>(*rootDevice, nullptr);
auto clSubDevice = std::make_unique<ClDevice>(*subDevice, *clRootDevice, nullptr);
auto clSubSubDevice = std::make_unique<ClDevice>(*subSubDevice, *clRootDevice, nullptr);
EXPECT_EQ(clRootDevice->getSubDevice(0), clRootDevice->getNearestGenericSubDevice(0));
EXPECT_EQ(clSubDevice.get(), clSubDevice->getNearestGenericSubDevice(0));
EXPECT_EQ(clRootDevice->getSubDevice(0), clSubSubDevice->getNearestGenericSubDevice(0));
}
TEST_F(EngineInstancedDeviceTests, givenAffinityMaskSetWhenCreatingDevicesThenFilterMaskedDevices) {
TEST_F(DeviceTests, givenAffinityMaskSetWhenCreatingDevicesThenFilterMaskedDevices) {
constexpr uint32_t genericDevicesCount = 3;
constexpr uint32_t ccsCount = 4;
constexpr uint32_t engineInstancedPerGeneric[3] = {3, 0, 2};
constexpr bool supportedGenericSubDevices[3] = {true, false, true};
constexpr bool supportedEngineDevices[3][4] = {{true, true, true, false},
{false, false, false, false},
{false, false, true, true}};
debugManager.flags.EngineInstancedSubDevices.set(true);
debugManager.flags.ZE_AFFINITY_MASK.set("0.0.0, 0.0.1, 0.0.2, 0.2.2, 0.2.3, 0.1.5");
if (!createDevices(genericDevicesCount, ccsCount)) {
@@ -754,35 +579,18 @@ TEST_F(EngineInstancedDeviceTests, givenAffinityMaskSetWhenCreatingDevicesThenFi
ASSERT_NE(nullptr, subDevice);
EXPECT_FALSE(subDevice->allEngines[0].osContext->isRootDevice());
EXPECT_FALSE(subDevice->engineInstanced);
EXPECT_EQ(engineInstancedPerGeneric[i], subDevice->getNumSubDevices());
EXPECT_EQ(0u, subDevice->getNumSubDevices());
EXPECT_EQ(0u, subDevice->getNumGenericSubDevices());
EXPECT_EQ(aub_stream::EngineType::NUM_ENGINES, subDevice->engineInstancedType);
EXPECT_TRUE(hasAllEngines(subDevice));
for (uint32_t j = 0; j < ccsCount; j++) {
if (!supportedEngineDevices[i][j]) {
EXPECT_EQ(nullptr, subDevice->getSubDevice(j));
continue;
}
auto engineType = static_cast<aub_stream::EngineType>(aub_stream::EngineType::ENGINE_CCS + j);
auto engineSubDevice = static_cast<MockSubDevice *>(subDevice->getSubDevice(j));
ASSERT_NE(nullptr, engineSubDevice);
EXPECT_TRUE(isEngineInstanced(engineSubDevice, engineType, subDevice->getSubDeviceIndex(), subDevice->getDeviceBitfield()));
EXPECT_TRUE(hasEngineInstancedEngines(engineSubDevice, engineType));
}
}
}
TEST_F(EngineInstancedDeviceTests, givenAffinityMaskForSingle3rdLevelDeviceWhenCreatingDevicesThenCreate2ndLevelAsEngineInstanced) {
TEST_F(DeviceTests, givenAffinityMaskForSingle3rdLevelDeviceWhenCreatingDevicesThenCreate2ndLevel) {
constexpr uint32_t genericDevicesCount = 2;
constexpr uint32_t ccsCount = 2;
constexpr uint32_t create2ndLevelAsEngineInstanced[2] = {false, true};
constexpr uint32_t engineInstanced2ndLevelEngineIndex = 1;
constexpr uint32_t create2ndLevel[2] = {false, true};
debugManager.flags.EngineInstancedSubDevices.set(true);
debugManager.flags.ZE_AFFINITY_MASK.set("0.0, 0.1.1");
if (!createDevices(genericDevicesCount, ccsCount)) {
@@ -798,13 +606,7 @@ TEST_F(EngineInstancedDeviceTests, givenAffinityMaskForSingle3rdLevelDeviceWhenC
EXPECT_FALSE(subDevice->allEngines[0].osContext->isRootDevice());
if (create2ndLevelAsEngineInstanced[i]) {
auto engineType = static_cast<aub_stream::EngineType>(aub_stream::EngineType::ENGINE_CCS + engineInstanced2ndLevelEngineIndex);
DeviceBitfield deviceBitfield = (1llu << i);
EXPECT_TRUE(isEngineInstanced(subDevice, engineType, i, deviceBitfield));
EXPECT_TRUE(hasEngineInstancedEngines(subDevice, engineType));
if (create2ndLevel[i]) {
EXPECT_EQ(0u, subDevice->getNumGenericSubDevices());
EXPECT_EQ(0u, subDevice->getNumSubDevices());
@@ -813,122 +615,12 @@ TEST_F(EngineInstancedDeviceTests, givenAffinityMaskForSingle3rdLevelDeviceWhenC
EXPECT_TRUE(hasAllEngines(subDevice));
EXPECT_FALSE(subDevice->engineInstanced);
EXPECT_EQ(aub_stream::EngineType::NUM_ENGINES, subDevice->engineInstancedType);
EXPECT_EQ(0u, subDevice->getNumGenericSubDevices());
EXPECT_EQ(ccsCount, subDevice->getNumSubDevices());
for (uint32_t j = 0; j < ccsCount; j++) {
auto engineType = static_cast<aub_stream::EngineType>(aub_stream::EngineType::ENGINE_CCS + j);
auto engineSubDevice = static_cast<MockSubDevice *>(subDevice->getSubDevice(j));
ASSERT_NE(nullptr, engineSubDevice);
EXPECT_TRUE(isEngineInstanced(engineSubDevice, engineType, subDevice->getSubDeviceIndex(), subDevice->getDeviceBitfield()));
EXPECT_TRUE(hasEngineInstancedEngines(engineSubDevice, engineType));
}
EXPECT_EQ(0u, subDevice->getNumSubDevices());
}
}
TEST_F(EngineInstancedDeviceTests, givenAffinityMaskForSingle3rdLevelDeviceOnlyWhenCreatingDevicesThenCreate1stLevelAsEngineInstanced) {
constexpr uint32_t genericDevicesCount = 2;
constexpr uint32_t ccsCount = 2;
constexpr uint32_t genericDeviceIndex = 1;
constexpr uint32_t engineInstancedEngineIndex = 1;
debugManager.flags.ZE_AFFINITY_MASK.set("0.1.1");
if (!createDevices(genericDevicesCount, ccsCount)) {
GTEST_SKIP();
}
EXPECT_FALSE(hasRootCsrOnly(rootDevice));
auto engineType = static_cast<aub_stream::EngineType>(aub_stream::EngineType::ENGINE_CCS + engineInstancedEngineIndex);
DeviceBitfield deviceBitfield = (1llu << genericDeviceIndex);
EXPECT_FALSE(rootDevice->allEngines[0].osContext->isRootDevice());
EXPECT_TRUE(rootDevice->engineInstanced);
EXPECT_TRUE(rootDevice->getNumGenericSubDevices() == 0);
EXPECT_TRUE(rootDevice->getNumSubDevices() == 0);
EXPECT_TRUE(engineType == rootDevice->engineInstancedType);
EXPECT_TRUE(deviceBitfield == rootDevice->getDeviceBitfield());
EXPECT_EQ(1u, rootDevice->getDeviceBitfield().count());
EXPECT_TRUE(hasEngineInstancedEngines(rootDevice, engineType));
}
TEST_F(EngineInstancedDeviceTests, givenAffinityMaskForSingle2rdLevelDeviceOnlyWhenCreatingDevicesThenCreate1stLevelAsEngineInstanced) {
constexpr uint32_t genericDevicesCount = 1;
constexpr uint32_t ccsCount = 2;
constexpr uint32_t genericDeviceIndex = 0;
constexpr uint32_t engineInstancedEngineIndex = 1;
debugManager.flags.ZE_AFFINITY_MASK.set("0.0.1, 0.9");
if (!createDevices(genericDevicesCount, ccsCount)) {
GTEST_SKIP();
}
EXPECT_FALSE(hasRootCsrOnly(rootDevice));
auto engineType = static_cast<aub_stream::EngineType>(aub_stream::EngineType::ENGINE_CCS + engineInstancedEngineIndex);
DeviceBitfield deviceBitfield = (1llu << genericDeviceIndex);
EXPECT_FALSE(rootDevice->allEngines[0].osContext->isRootDevice());
EXPECT_TRUE(rootDevice->engineInstanced);
EXPECT_TRUE(rootDevice->getNumGenericSubDevices() == 0);
EXPECT_TRUE(rootDevice->getNumSubDevices() == 0);
EXPECT_TRUE(engineType == rootDevice->engineInstancedType);
EXPECT_TRUE(deviceBitfield == rootDevice->getDeviceBitfield());
EXPECT_EQ(1u, rootDevice->getDeviceBitfield().count());
EXPECT_TRUE(hasEngineInstancedEngines(rootDevice, engineType));
}
TEST_F(EngineInstancedDeviceTests, givenAffinityMaskForSecondLevelOnSingleTileDeviceWhenCreatingThenEnableAllEngineInstancedDevices) {
constexpr uint32_t genericDevicesCount = 1;
constexpr uint32_t ccsCount = 2;
debugManager.flags.EngineInstancedSubDevices.set(true);
debugManager.flags.AllowSingleTileEngineInstancedSubDevices.set(true);
debugManager.flags.ZE_AFFINITY_MASK.set("0.0, 0.4");
if (!createDevices(genericDevicesCount, ccsCount)) {
GTEST_SKIP();
}
EXPECT_FALSE(hasRootCsrOnly(rootDevice));
EXPECT_TRUE(rootDevice->isEngineInstanced());
EXPECT_EQ(0u, rootDevice->getNumGenericSubDevices());
EXPECT_EQ(0u, rootDevice->getNumSubDevices());
}
TEST_F(EngineInstancedDeviceTests, givenAffinityMaskForSecondLevelOnSingleTileDeviceSingleEngineWhenCreatingThenDontEnableEngineInstancedDevices) {
constexpr uint32_t genericDevicesCount = 1;
constexpr uint32_t ccsCount = 1;
debugManager.flags.EngineInstancedSubDevices.set(true);
debugManager.flags.AllowSingleTileEngineInstancedSubDevices.set(true);
debugManager.flags.ZE_AFFINITY_MASK.set("0.0");
if (!createDevices(genericDevicesCount, ccsCount)) {
GTEST_SKIP();
}
EXPECT_FALSE(hasRootCsrOnly(rootDevice));
EXPECT_FALSE(rootDevice->isEngineInstanced());
EXPECT_EQ(0u, rootDevice->getNumGenericSubDevices());
EXPECT_EQ(0u, rootDevice->getNumSubDevices());
}
TEST_F(EngineInstancedDeviceTests, givenAffinityMaskForSecondLevelOnSingleTileDeviceWithoutDebugFlagWhenCreatingThenDontEnableAllEngineInstancedDevices) {
TEST_F(DeviceTests, givenAffinityMaskForSecondLevelOnSingleTileDeviceWithoutDebugFlagWhenCreatingThenDontEnableAllSubDevices) {
constexpr uint32_t genericDevicesCount = 1;
constexpr uint32_t ccsCount = 2;
@@ -940,12 +632,11 @@ TEST_F(EngineInstancedDeviceTests, givenAffinityMaskForSecondLevelOnSingleTileDe
EXPECT_FALSE(hasRootCsrOnly(rootDevice));
EXPECT_FALSE(rootDevice->isEngineInstanced());
EXPECT_EQ(0u, rootDevice->getNumGenericSubDevices());
EXPECT_EQ(0u, rootDevice->getNumSubDevices());
}
TEST_F(EngineInstancedDeviceTests, givenAffinityMaskWhenCreatingClSubDevicesThenSkipDisabledDevices) {
TEST_F(DeviceTests, givenAffinityMaskWhenCreatingClSubDevicesThenSkipDisabledDevices) {
constexpr uint32_t genericDevicesCount = 3;
constexpr uint32_t ccsCount = 1;
@@ -970,78 +661,7 @@ struct SingleSliceDispatchSupportMatcher {
}
};
HWTEST2_F(EngineInstancedDeviceTests, givenEngineInstancedDeviceWhenProgrammingCfeStateThenSetSingleSliceDispatch, SingleSliceDispatchSupportMatcher) {
using CFE_STATE = typename FamilyType::CFE_STATE;
debugManager.flags.EngineInstancedSubDevices.set(true);
constexpr uint32_t genericDevicesCount = 1;
constexpr uint32_t ccsCount = 2;
debugManager.flags.AllowSingleTileEngineInstancedSubDevices.set(true);
if (!createDevices(genericDevicesCount, ccsCount)) {
GTEST_SKIP();
}
auto subDevice = static_cast<MockSubDevice *>(rootDevice->getSubDevice(0));
auto defaultEngine = subDevice->getDefaultEngine();
EXPECT_TRUE(defaultEngine.osContext->isEngineInstanced());
char buffer[64] = {};
MockGraphicsAllocation graphicsAllocation(buffer, sizeof(buffer));
LinearStream linearStream(&graphicsAllocation, graphicsAllocation.getUnderlyingBuffer(), graphicsAllocation.getUnderlyingBufferSize());
auto csr = static_cast<UltCommandStreamReceiver<FamilyType> *>(defaultEngine.commandStreamReceiver);
auto dispatchFlags = DispatchFlagsHelper::createDefaultDispatchFlags();
csr->programVFEState(linearStream, dispatchFlags, 1);
auto cfeState = reinterpret_cast<CFE_STATE *>(buffer);
EXPECT_TRUE(cfeState->getSingleSliceDispatchCcsMode());
}
HWTEST_F(EngineInstancedDeviceTests, givenEngineInstancedDeviceWhenCreatingProgramThenAssignAllSubDevices) {
constexpr uint32_t genericDevicesCount = 2;
constexpr uint32_t ccsCount = 2;
debugManager.flags.EngineInstancedSubDevices.set(true);
if (!createDevices(genericDevicesCount, ccsCount)) {
GTEST_SKIP();
}
const char *source = "text";
size_t sourceSize = strlen(source);
auto clRootDevice = std::make_unique<ClDevice>(*rootDevice, nullptr);
auto clSubDevice = clRootDevice->getSubDevice(0);
auto clSubSubDevice0 = clSubDevice->getSubDevice(0);
auto clSubSubDevice1 = clSubDevice->getSubDevice(1);
cl_device_id deviceIds[] = {clSubDevice, clSubSubDevice0, clSubSubDevice1};
ClDeviceVector deviceVector{deviceIds, 3};
MockContext context(deviceVector);
cl_int retVal = CL_INVALID_PROGRAM;
auto program = std::unique_ptr<MockProgram>(Program::create<MockProgram>(
&context,
1,
&source,
&sourceSize,
retVal));
ASSERT_NE(nullptr, program.get());
ASSERT_EQ(CL_SUCCESS, retVal);
ASSERT_TRUE(program->deviceBuildInfos.find(clSubDevice) != program->deviceBuildInfos.end());
auto &associatedSubDevices = program->deviceBuildInfos[clSubDevice].associatedSubDevices;
ASSERT_EQ(2u, associatedSubDevices.size());
EXPECT_EQ(clSubSubDevice0, associatedSubDevices[0]);
EXPECT_EQ(clSubSubDevice1, associatedSubDevices[1]);
}
HWTEST_F(EngineInstancedDeviceTests, whenCreateMultipleCommandQueuesThenEnginesAreAssignedUsingRoundRobin) {
HWTEST_F(DeviceTests, whenCreateMultipleCommandQueuesThenEnginesAreAssignedUsingRoundRobin) {
constexpr uint32_t genericDevicesCount = 1;
constexpr uint32_t ccsCount = 4;
@@ -1087,7 +707,7 @@ HWTEST_F(EngineInstancedDeviceTests, whenCreateMultipleCommandQueuesThenEnginesA
}
}
HWTEST_F(EngineInstancedDeviceTests, givenCmdQRoundRobindEngineAssignBitfieldwWenCreateMultipleCommandQueuesThenEnginesAreAssignedUsingRoundRobinSkippingNotAvailableEngines) {
HWTEST_F(DeviceTests, givenCmdQRoundRobindEngineAssignBitfieldwWenCreateMultipleCommandQueuesThenEnginesAreAssignedUsingRoundRobinSkippingNotAvailableEngines) {
constexpr uint32_t genericDevicesCount = 1;
constexpr uint32_t ccsCount = 4;
@@ -1137,7 +757,7 @@ HWTEST_F(EngineInstancedDeviceTests, givenCmdQRoundRobindEngineAssignBitfieldwWe
}
}
HWTEST_F(EngineInstancedDeviceTests, givenCmdQRoundRobindEngineAssignNTo1wWenCreateMultipleCommandQueuesThenEnginesAreAssignedUsingRoundRobinAndNQueuesShareSameCsr) {
HWTEST_F(DeviceTests, givenCmdQRoundRobindEngineAssignNTo1wWenCreateMultipleCommandQueuesThenEnginesAreAssignedUsingRoundRobinAndNQueuesShareSameCsr) {
constexpr uint32_t genericDevicesCount = 1;
constexpr uint32_t ccsCount = 4;
@@ -1184,7 +804,7 @@ HWTEST_F(EngineInstancedDeviceTests, givenCmdQRoundRobindEngineAssignNTo1wWenCre
}
}
HWTEST_F(EngineInstancedDeviceTests, givenCmdQRoundRobindEngineAssignNTo1AndCmdQRoundRobindEngineAssignBitfieldwWenCreateMultipleCommandQueuesThenEnginesAreAssignedProperlyUsingRoundRobin) {
HWTEST_F(DeviceTests, givenCmdQRoundRobindEngineAssignNTo1AndCmdQRoundRobindEngineAssignBitfieldwWenCreateMultipleCommandQueuesThenEnginesAreAssignedProperlyUsingRoundRobin) {
constexpr uint32_t genericDevicesCount = 1;
constexpr uint32_t ccsCount = 4;
@@ -1235,7 +855,7 @@ HWTEST_F(EngineInstancedDeviceTests, givenCmdQRoundRobindEngineAssignNTo1AndCmdQ
}
}
HWTEST_F(EngineInstancedDeviceTests, givenEnableCmdQRoundRobindEngineAssignDisabledWenCreateMultipleCommandQueuesThenDefaultEngineAssigned) {
HWTEST_F(DeviceTests, givenEnableCmdQRoundRobindEngineAssignDisabledWenCreateMultipleCommandQueuesThenDefaultEngineAssigned) {
constexpr uint32_t genericDevicesCount = 1;
constexpr uint32_t ccsCount = 4;
@@ -1312,8 +932,5 @@ TEST(SubDevicesTest, givenCreateMultipleSubDevicesFlagSetToOneWhenCreateRootDevi
ultHwConfig.useMockedPrepareDeviceEnvironmentsFunc = false;
initPlatform();
auto rootDevice = static_cast<RootDevice *>(&platform()->getClDevice(0)->getDevice());
if (rootDevice->getNumSubDevices() > 0) {
EXPECT_TRUE(rootDevice->getSubDevice(0)->isEngineInstanced());
}
EXPECT_EQ(0u, rootDevice->getNumGenericSubDevices());
}

2
opencl/test/unit_test/fixtures/ult_command_stream_receiver_fixture.h
View File

@@ -190,7 +190,7 @@ struct UltCommandStreamReceiverTest
commandStreamReceiver.streamProperties.pipelineSelect.setPropertiesAll(true, false, false);
commandStreamReceiver.streamProperties.stateComputeMode.setPropertiesAll(0, GrfConfig::defaultGrfNumber,
gfxCoreHelper.getDefaultThreadArbitrationPolicy(), pDevice->getPreemptionMode());
commandStreamReceiver.streamProperties.frontEndState.setPropertiesAll(false, false, false, -1);
commandStreamReceiver.streamProperties.frontEndState.setPropertiesAll(false, false, false);
}
template <typename GfxFamily>

2
opencl/test/unit_test/mem_obj/buffer_bcs_tests.cpp
View File

@@ -614,7 +614,7 @@ HWTEST_TEMPLATED_F(BcsBufferTests, givenAllBcsEnginesReadyWhenWaitingForEventThe
EXPECT_EQ(copyDefaultEngineType, eventObj->getBcsEngineType());
MockOsContext osContext(123, {{aub_stream::EngineType::ENGINE_BCS2, EngineUsage::regular}, device->getDeviceBitfield(), PreemptionMode::Disabled, false, false});
MockOsContext osContext(123, {{aub_stream::EngineType::ENGINE_BCS2, EngineUsage::regular}, device->getDeviceBitfield(), PreemptionMode::Disabled, false});
UltCommandStreamReceiver<FamilyType> ultCsr2(*device->getExecutionEnvironment(), 0, device->getDeviceBitfield());
ultCsr2.initializeTagAllocation();
ultCsr2.setupContext(osContext);

4
opencl/test/unit_test/xe2_hpg_core/test_preamble_xe2_hpg_core.cpp
View File

@@ -18,7 +18,7 @@ XE2_HPG_CORETEST_F(PreambleCfeState, givenXe2HpgCoreAndConcurrentKernelExecution
auto pVfeCmd = PreambleHelper<FamilyType>::getSpaceForVfeState(&linearStream, *defaultHwInfo, EngineGroupType::renderCompute);
StreamProperties streamProperties{};
streamProperties.initSupport(pDevice->getRootDeviceEnvironment());
streamProperties.frontEndState.setPropertiesAll(true, false, false, false);
streamProperties.frontEndState.setPropertiesAll(true, false, false);
PreambleHelper<FamilyType>::programVfeState(pVfeCmd, pDevice->getRootDeviceEnvironment(), 0u, 0, 0, streamProperties);
parseCommands<FamilyType>(linearStream);
auto cfeStateIt = find<CFE_STATE *>(cmdList.begin(), cmdList.end());
@@ -34,7 +34,7 @@ XE2_HPG_CORETEST_F(PreambleCfeState, givenXe2HpgCoreAndDefaultKernelExecutionTyp
auto pVfeCmd = PreambleHelper<FamilyType>::getSpaceForVfeState(&linearStream, *defaultHwInfo, EngineGroupType::renderCompute);
StreamProperties streamProperties{};
streamProperties.initSupport(pDevice->getRootDeviceEnvironment());
streamProperties.frontEndState.setPropertiesAll(false, false, false, false);
streamProperties.frontEndState.setPropertiesAll(false, false, false);
PreambleHelper<FamilyType>::programVfeState(pVfeCmd, pDevice->getRootDeviceEnvironment(), 0u, 0, 0, streamProperties);
parseCommands<FamilyType>(linearStream);
auto cfeStateIt = find<CFE_STATE *>(cmdList.begin(), cmdList.end());

6
opencl/test/unit_test/xe_hpg_core/test_cmds_programming_xe_hpg_core.cpp
View File

@@ -336,7 +336,7 @@ HWTEST2_F(PreambleCfeState, givenXehpAndDisabledFusedEuWhenCfeStateProgrammedThe
auto pVfeCmd = PreambleHelper<FamilyType>::getSpaceForVfeState(&linearStream, hwInfo, EngineGroupType::renderCompute);
StreamProperties streamProperties{};
streamProperties.initSupport(rootDeviceEnvironment);
streamProperties.frontEndState.setPropertiesAll(false, false, false, false);
streamProperties.frontEndState.setPropertiesAll(false, false, false);
PreambleHelper<FamilyType>::programVfeState(pVfeCmd, rootDeviceEnvironment, 0u, 0, 0, streamProperties);
parseCommands<FamilyType>(linearStream);
auto cfeStateIt = find<CFE_STATE *>(cmdList.begin(), cmdList.end());
@@ -359,7 +359,7 @@ HWTEST2_F(PreambleCfeState, givenXehpEnabledFusedEuAndDisableFusedDispatchFromKe
auto pVfeCmd = PreambleHelper<FamilyType>::getSpaceForVfeState(&linearStream, hwInfo, EngineGroupType::renderCompute);
StreamProperties streamProperties{};
streamProperties.initSupport(rootDeviceEnvironment);
streamProperties.frontEndState.setPropertiesAll(false, true, false, false);
streamProperties.frontEndState.setPropertiesAll(false, true, false);
PreambleHelper<FamilyType>::programVfeState(pVfeCmd, rootDeviceEnvironment, 0u, 0, 0, streamProperties);
parseCommands<FamilyType>(linearStream);
auto cfeStateIt = find<CFE_STATE *>(cmdList.begin(), cmdList.end());
@@ -379,7 +379,7 @@ HWTEST2_F(PreambleCfeState, givenXehpAndEnabledFusedEuWhenCfeStateProgrammedThen
auto pVfeCmd = PreambleHelper<FamilyType>::getSpaceForVfeState(&linearStream, hwInfo, EngineGroupType::renderCompute);
StreamProperties streamProperties{};
streamProperties.initSupport(rootDeviceEnvironment);
streamProperties.frontEndState.setPropertiesAll(false, false, false, false);
streamProperties.frontEndState.setPropertiesAll(false, false, false);
PreambleHelper<FamilyType>::programVfeState(pVfeCmd, rootDeviceEnvironment, 0u, 0, 0, streamProperties);
parseCommands<FamilyType>(linearStream);
auto cfeStateIt = find<CFE_STATE *>(cmdList.begin(), cmdList.end());

2
shared/source/command_stream/command_stream_receiver_hw_base.inl
View File

@@ -885,7 +885,7 @@ inline void CommandStreamReceiverHw<GfxFamily>::programVFEState(LinearStream &cs
auto isCooperative = dispatchFlags.kernelExecutionType == KernelExecutionType::concurrent;
auto disableOverdispatch = (dispatchFlags.additionalKernelExecInfo != AdditionalKernelExecInfo::notSet);
this->streamProperties.frontEndState.setPropertiesAll(isCooperative, dispatchFlags.disableEUFusion, disableOverdispatch, osContext->isEngineInstanced());
this->streamProperties.frontEndState.setPropertiesAll(isCooperative, dispatchFlags.disableEUFusion, disableOverdispatch);
auto &gfxCoreHelper = getGfxCoreHelper();
auto engineGroupType = gfxCoreHelper.getEngineGroupType(getOsContext().getEngineType(), getOsContext().getEngineUsage(), hwInfo);

6
shared/source/command_stream/definitions/stream_properties.inl
View File

@@ -79,9 +79,9 @@ struct FrontEndProperties {
void initSupport(const RootDeviceEnvironment &rootDeviceEnvironment);
void resetState();
void setPropertiesAll(bool isCooperativeKernel, bool disableEuFusion, bool disableOverdispatch, int32_t engineInstancedDevice);
void setPropertySingleSliceDispatchCcsMode(int32_t engineInstancedDevice);
void setPropertiesDisableOverdispatchEngineInstanced(bool disableOverdispatch, int32_t engineInstancedDevice, bool clearDirtyState);
void setPropertiesAll(bool isCooperativeKernel, bool disableEuFusion, bool disableOverdispatch);
void setPropertySingleSliceDispatchCcsMode();
void setPropertiesDisableOverdispatch(bool disableOverdispatch, bool clearDirtyState);
void setPropertiesComputeDispatchAllWalkerEnableDisableEuFusion(bool isCooperativeKernel, bool disableEuFusion);
void copyPropertiesAll(const FrontEndProperties &properties);

16
shared/source/command_stream/stream_properties.cpp
View File

@@ -196,7 +196,7 @@ void FrontEndProperties::resetState() {
this->singleSliceDispatchCcsMode.value = StreamProperty::initValue;
}
void FrontEndProperties::setPropertiesAll(bool isCooperativeKernel, bool disableEuFusion, bool disableOverdispatch, int32_t engineInstancedDevice) {
void FrontEndProperties::setPropertiesAll(bool isCooperativeKernel, bool disableEuFusion, bool disableOverdispatch) {
DEBUG_BREAK_IF(!this->propertiesSupportLoaded);
clearIsDirty();
@@ -211,22 +211,15 @@ void FrontEndProperties::setPropertiesAll(bool isCooperativeKernel, bool disable
if (this->frontEndPropertiesSupport.disableOverdispatch) {
this->disableOverdispatch.set(disableOverdispatch);
}
if (this->frontEndPropertiesSupport.singleSliceDispatchCcsMode) {
this->singleSliceDispatchCcsMode.set(engineInstancedDevice);
}
}
void FrontEndProperties::setPropertySingleSliceDispatchCcsMode(int32_t engineInstancedDevice) {
void FrontEndProperties::setPropertySingleSliceDispatchCcsMode() {
DEBUG_BREAK_IF(!this->propertiesSupportLoaded);
this->singleSliceDispatchCcsMode.isDirty = false;
if (this->frontEndPropertiesSupport.singleSliceDispatchCcsMode) {
this->singleSliceDispatchCcsMode.set(engineInstancedDevice);
}
}
void FrontEndProperties::setPropertiesDisableOverdispatchEngineInstanced(bool disableOverdispatch, int32_t engineInstancedDevice, bool clearDirtyState) {
void FrontEndProperties::setPropertiesDisableOverdispatch(bool disableOverdispatch, bool clearDirtyState) {
DEBUG_BREAK_IF(!this->propertiesSupportLoaded);
if (!clearDirtyState) {
@@ -237,9 +230,6 @@ void FrontEndProperties::setPropertiesDisableOverdispatchEngineInstanced(bool di
if (this->frontEndPropertiesSupport.disableOverdispatch) {
this->disableOverdispatch.set(disableOverdispatch);
}
if (this->frontEndPropertiesSupport.singleSliceDispatchCcsMode) {
this->singleSliceDispatchCcsMode.set(engineInstancedDevice);
}
if (clearDirtyState) {
this->disableOverdispatch.isDirty = false;

2
shared/source/debug_settings/debug_variables_base.inl
View File

@@ -457,8 +457,6 @@ DECLARE_DEBUG_VARIABLE(bool, EnableExtendedVaFormats, false, "Enable more format
DECLARE_DEBUG_VARIABLE(bool, EnableFormatQuery, true, "Enable sharing format querying")
DECLARE_DEBUG_VARIABLE(bool, EnableFreeMemory, true, "Enable freeMemory in memory manager")
DECLARE_DEBUG_VARIABLE(bool, ForceSamplerLowFilteringPrecision, false, "Force Low Filtering Precision Sampler mode")
DECLARE_DEBUG_VARIABLE(bool, EngineInstancedSubDevices, false, "Create subdevices assigned to specific engine")
DECLARE_DEBUG_VARIABLE(bool, AllowSingleTileEngineInstancedSubDevices, false, "Create subdevices assigned to specific engine on single tile config")
DECLARE_DEBUG_VARIABLE(bool, EnablePrivateBO, false, "Enable PRELIM_I915_GEM_CREATE_EXT_VM_PRIVATE extension creating VM_PRIVATE BOs")
DECLARE_DEBUG_VARIABLE(bool, EnableAIL, true, "Enables AIL")
DECLARE_DEBUG_VARIABLE(int64_t, VmBindWaitUserFenceTimeout, -1, "-1: default, >0: time in ns for wait function timeout")

146
shared/source/device/device.cpp
View File

@@ -83,10 +83,6 @@ SubDevice *Device::createSubDevice(uint32_t subDeviceIndex) {
return Device::create<SubDevice>(executionEnvironment, subDeviceIndex, *getRootDevice());
}
SubDevice *Device::createEngineInstancedSubDevice(uint32_t subDeviceIndex, aub_stream::EngineType engineType) {
return Device::create<SubDevice>(executionEnvironment, subDeviceIndex, *getRootDevice(), engineType);
}
bool Device::genericSubDevicesAllowed() {
auto deviceMask = executionEnvironment->rootDeviceEnvironments[getRootDeviceIndex()]->deviceAffinityMask.getGenericSubDevicesMask();
uint32_t subDeviceCount = GfxCoreHelper::getSubDevicesCount(&getHardwareInfo());
@@ -100,56 +96,6 @@ bool Device::genericSubDevicesAllowed() {
return (numSubDevices > 0);
}
bool Device::engineInstancedSubDevicesAllowed() {
bool notAllowed = !debugManager.flags.EngineInstancedSubDevices.get();
notAllowed |= engineInstanced;
notAllowed |= (getHardwareInfo().gtSystemInfo.CCSInfo.NumberOfCCSEnabled < 2);
notAllowed |= ((GfxCoreHelper::getSubDevicesCount(&getHardwareInfo()) < 2) && (!debugManager.flags.AllowSingleTileEngineInstancedSubDevices.get()));
if (notAllowed) {
return false;
}
UNRECOVERABLE_IF(deviceBitfield.count() != 1);
uint32_t subDeviceIndex = Math::log2(static_cast<uint32_t>(deviceBitfield.to_ulong()));
auto enginesMask = getRootDeviceEnvironment().deviceAffinityMask.getEnginesMask(subDeviceIndex);
auto ccsCount = getHardwareInfo().gtSystemInfo.CCSInfo.NumberOfCCSEnabled;
numSubDevices = std::min(ccsCount, static_cast<uint32_t>(enginesMask.count()));
if (numSubDevices == 1) {
numSubDevices = 0;
}
return (numSubDevices > 0);
}
bool Device::createEngineInstancedSubDevices() {
UNRECOVERABLE_IF(deviceBitfield.count() != 1);
UNRECOVERABLE_IF(!subdevices.empty());
uint32_t subDeviceIndex = Math::log2(static_cast<uint32_t>(deviceBitfield.to_ulong()));
auto enginesMask = getRootDeviceEnvironment().deviceAffinityMask.getEnginesMask(subDeviceIndex);
auto ccsCount = getHardwareInfo().gtSystemInfo.CCSInfo.NumberOfCCSEnabled;
subdevices.resize(ccsCount, nullptr);
for (uint32_t i = 0; i < ccsCount; i++) {
if (!enginesMask.test(i)) {
continue;
}
auto engineType = static_cast<aub_stream::EngineType>(aub_stream::EngineType::ENGINE_CCS + i);
auto subDevice = createEngineInstancedSubDevice(subDeviceIndex, engineType);
UNRECOVERABLE_IF(!subDevice);
subdevices[i] = subDevice;
}
return true;
}
bool Device::createGenericSubDevices() {
UNRECOVERABLE_IF(!subdevices.empty());
uint32_t subDeviceCount = GfxCoreHelper::getSubDevicesCount(&getHardwareInfo());
@@ -176,42 +122,9 @@ bool Device::createSubDevices() {
return createGenericSubDevices();
}
if (engineInstancedSubDevicesAllowed()) {
return createEngineInstancedSubDevices();
}
return true;
}
void Device::setAsEngineInstanced() {
if (subdevices.size() > 0) {
return;
}
UNRECOVERABLE_IF(deviceBitfield.count() != 1);
uint32_t subDeviceIndex = Math::log2(static_cast<uint32_t>(deviceBitfield.to_ulong()));
auto enginesMask = getRootDeviceEnvironment().deviceAffinityMask.getEnginesMask(subDeviceIndex);
if (enginesMask.count() != 1) {
return;
}
auto ccsCount = getHardwareInfo().gtSystemInfo.CCSInfo.NumberOfCCSEnabled;
for (uint32_t i = 0; i < ccsCount; i++) {
if (!enginesMask.test(i)) {
continue;
}
UNRECOVERABLE_IF(engineInstanced);
engineInstanced = true;
engineInstancedType = static_cast<aub_stream::EngineType>(aub_stream::EngineType::ENGINE_CCS + i);
}
UNRECOVERABLE_IF(!engineInstanced);
}
bool Device::createDeviceImpl() {
// init sub devices first
if (!createSubDevices()) {
@@ -243,8 +156,6 @@ bool Device::createDeviceImpl() {
}
bool Device::initDeviceWithEngines() {
setAsEngineInstanced();
auto &productHelper = getProductHelper();
if (getDebugger() && productHelper.disableL3CacheForDebug(getHardwareInfo())) {
getGmmHelper()->forceAllResourcesUncached();
@@ -338,37 +249,31 @@ bool Device::initDeviceFully() {
}
createBindlessHeapsHelper();
if (!isEngineInstanced()) {
uuid.isValid = false;
uuid.isValid = false;
if (getRootDeviceEnvironment().osInterface == nullptr) {
return true;
if (getRootDeviceEnvironment().osInterface == nullptr) {
return true;
}
auto &gfxCoreHelper = getGfxCoreHelper();
auto &productHelper = getProductHelper();
if (debugManager.flags.EnableChipsetUniqueUUID.get() != 0) {
if (gfxCoreHelper.isChipsetUniqueUUIDSupported()) {
auto deviceIndex = isSubDevice() ? static_cast<SubDevice *>(this)->getSubDeviceIndex() + 1 : 0;
uuid.isValid = productHelper.getUuid(getRootDeviceEnvironment().osInterface->getDriverModel(), getRootDevice()->getNumSubDevices(), deviceIndex, uuid.id);
}
}
auto &gfxCoreHelper = getGfxCoreHelper();
auto &productHelper = getProductHelper();
if (debugManager.flags.EnableChipsetUniqueUUID.get() != 0) {
if (gfxCoreHelper.isChipsetUniqueUUIDSupported()) {
auto deviceIndex = isSubDevice() ? static_cast<SubDevice *>(this)->getSubDeviceIndex() + 1 : 0;
uuid.isValid = productHelper.getUuid(getRootDeviceEnvironment().osInterface->getDriverModel(), getRootDevice()->getNumSubDevices(), deviceIndex, uuid.id);
}
}
if (!uuid.isValid) {
PhysicalDevicePciBusInfo pciBusInfo = getRootDeviceEnvironment().osInterface->getDriverModel()->getPciBusInfo();
uuid.isValid = generateUuidFromPciBusInfo(pciBusInfo, uuid.id);
}
if (!uuid.isValid) {
PhysicalDevicePciBusInfo pciBusInfo = getRootDeviceEnvironment().osInterface->getDriverModel()->getPciBusInfo();
uuid.isValid = generateUuidFromPciBusInfo(pciBusInfo, uuid.id);
}
return true;
}
bool Device::createEngines() {
if (engineInstanced) {
return createEngine({engineInstancedType, EngineUsage::regular});
}
auto &gfxCoreHelper = getGfxCoreHelper();
auto gpgpuEngines = gfxCoreHelper.getGpgpuEngineInstances(getRootDeviceEnvironment());
@@ -529,9 +434,8 @@ bool Device::createEngine(EngineTypeUsage engineTypeUsage) {
auto &gfxCoreHelper = getGfxCoreHelper();
const auto engineType = engineTypeUsage.first;
const auto engineUsage = engineTypeUsage.second;
const auto defaultEngineType = engineInstanced ? this->engineInstancedType : getChosenEngineType(hwInfo);
const auto defaultEngineType = getChosenEngineType(hwInfo);
const bool isDefaultEngine = defaultEngineType == engineType && engineUsage == EngineUsage::regular;
const bool createAsEngineInstanced = engineInstanced && EngineHelpers::isCcs(engineType);
bool primaryEngineTypeAllowed = (EngineHelpers::isCcs(engineType) || EngineHelpers::isBcs(engineType));
@@ -558,7 +462,7 @@ bool Device::createEngine(EngineTypeUsage engineTypeUsage) {
commandStreamReceiver->createPageTableManager();
}
EngineDescriptor engineDescriptor(engineTypeUsage, getDeviceBitfield(), preemptionMode, false, createAsEngineInstanced);
EngineDescriptor engineDescriptor(engineTypeUsage, getDeviceBitfield(), preemptionMode, false);
auto osContext = executionEnvironment->memoryManager->createAndRegisterOsContext(commandStreamReceiver.get(), engineDescriptor);
osContext->setContextGroup(useContextGroup);
@@ -648,7 +552,7 @@ bool Device::createSecondaryEngine(CommandStreamReceiver *primaryCsr, EngineType
commandStreamReceiver->initializeDefaultsForInternalEngine();
}
EngineDescriptor engineDescriptor(engineTypeUsage, getDeviceBitfield(), preemptionMode, false, false);
EngineDescriptor engineDescriptor(engineTypeUsage, getDeviceBitfield(), preemptionMode, false);
auto osContext = executionEnvironment->memoryManager->createAndRegisterSecondaryOsContext(&primaryCsr->getOsContext(), commandStreamReceiver.get(), engineDescriptor);
osContext->incRefInternal();
@@ -693,7 +597,7 @@ EngineControl *Device::getSecondaryEngineCsr(EngineTypeUsage engineTypeUsage, bo
commandStreamReceiver->createPageTableManager();
}
EngineDescriptor engineDescriptor(engineTypeUsage, getDeviceBitfield(), preemptionMode, false, false);
EngineDescriptor engineDescriptor(engineTypeUsage, getDeviceBitfield(), preemptionMode, false);
if (!commandStreamReceiver->initializeResources(allocateInterrupt)) {
return nullptr;
@@ -825,16 +729,6 @@ Device *Device::getSubDevice(uint32_t deviceId) const {
}
Device *Device::getNearestGenericSubDevice(uint32_t deviceId) {
/*
* EngineInstanced: Upper level
* Generic SubDevice: 'this'
* RootCsr Device: Next level SubDevice (generic)
*/
if (engineInstanced) {
return getRootDevice()->getNearestGenericSubDevice(Math::log2(static_cast<uint32_t>(deviceBitfield.to_ulong())));
}
if (subdevices.empty() || !hasRootCsr()) {
return this;
}

7
shared/source/device/device.h
View File

@@ -170,7 +170,6 @@ class Device : public ReferenceTrackedObject<Device> {
uint32_t getNumSubDevices() const { return numSubDevices; }
virtual bool isSubDevice() const = 0;
bool hasRootCsr() const { return rootCsrCreated; }
bool isEngineInstanced() const { return engineInstanced; }
BindlessHeapsHelper *getBindlessHeapsHelper() const;
@@ -245,16 +244,12 @@ class Device : public ReferenceTrackedObject<Device> {
MOCKABLE_VIRTUAL std::unique_ptr<CommandStreamReceiver> createCommandStreamReceiver() const;
MOCKABLE_VIRTUAL SubDevice *createSubDevice(uint32_t subDeviceIndex);
MOCKABLE_VIRTUAL SubDevice *createEngineInstancedSubDevice(uint32_t subDeviceIndex, aub_stream::EngineType engineType);
MOCKABLE_VIRTUAL size_t getMaxParameterSizeFromIGC() const;
double getPercentOfGlobalMemoryAvailable() const;
virtual void createBindlessHeapsHelper() {}
bool createSubDevices();
bool createGenericSubDevices();
bool createEngineInstancedSubDevices();
virtual bool genericSubDevicesAllowed();
bool engineInstancedSubDevicesAllowed();
void setAsEngineInstanced();
void finalizeRayTracing();
void createSecondaryContexts(const EngineControl &primaryEngine, SecondaryContexts &secondaryEnginesForType, uint32_t contextCount, uint32_t regularPriorityCount, uint32_t highPriorityContextCount);
void allocateDebugSurface(size_t debugSurfaceSize);
@@ -273,7 +268,6 @@ class Device : public ReferenceTrackedObject<Device> {
PreemptionMode preemptionMode = PreemptionMode::Disabled;
ExecutionEnvironment *executionEnvironment = nullptr;
aub_stream::EngineType engineInstancedType = aub_stream::EngineType::NUM_ENGINES;
uint32_t defaultEngineIndex = 0;
uint32_t numSubDevices = 0;
std::atomic_uint32_t regularCommandQueuesCreatedWithinDeviceCount{0};
@@ -281,7 +275,6 @@ class Device : public ReferenceTrackedObject<Device> {
uint32_t queuesPerEngineCount = 1;
void initializeEngineRoundRobinControls();
bool hasGenericSubDevices = false;
bool engineInstanced = false;
bool rootCsrCreated = false;
const uint32_t rootDeviceIndex;
GraphicsAllocation *debugSurface = nullptr;

2
shared/source/device/root_device.cpp
View File

@@ -72,7 +72,7 @@ void RootDevice::initializeRootCommandStreamReceiver() {
auto defaultEngineType = getChosenEngineType(hwInfo);
auto preemptionMode = PreemptionHelper::getDefaultPreemptionMode(hwInfo);
EngineDescriptor engineDescriptor(EngineTypeUsage{defaultEngineType, EngineUsage::regular}, getDeviceBitfield(), preemptionMode, true, false);
EngineDescriptor engineDescriptor(EngineTypeUsage{defaultEngineType, EngineUsage::regular}, getDeviceBitfield(), preemptionMode, true);
auto &gfxCoreHelper = getGfxCoreHelper();
bool isPrimaryEngine = EngineHelpers::isCcs(defaultEngineType);

4
shared/source/device/sub_device.cpp
View File

@@ -1,5 +1,5 @@
/*
* Copyright (C) 2019-2023 Intel Corporation
* Copyright (C) 2019-2024 Intel Corporation
*
* SPDX-License-Identifier: MIT
*
@@ -22,8 +22,6 @@ SubDevice::SubDevice(ExecutionEnvironment *executionEnvironment, uint32_t subDev
SubDevice::SubDevice(ExecutionEnvironment *executionEnvironment, uint32_t subDeviceIndex, Device &rootDevice, aub_stream::EngineType engineType)
: SubDevice(executionEnvironment, subDeviceIndex, rootDevice) {
this->engineInstancedType = engineType;
engineInstanced = true;
}
void SubDevice::incRefInternal() {

21
shared/source/execution_environment/execution_environment.cpp
View File

@@ -97,7 +97,6 @@ void ExecutionEnvironment::calculateMaxOsContextCount() {
auto &engineInstances = gfxCoreHelper.getGpgpuEngineInstances(*rootDeviceEnvironment);
auto osContextCount = static_cast<uint32_t>(engineInstances.size());
auto subDevicesCount = GfxCoreHelper::getSubDevicesCount(hwInfo);
auto ccsCount = hwInfo->gtSystemInfo.CCSInfo.NumberOfCCSEnabled;
bool hasRootCsr = subDevicesCount > 1;
uint32_t numRegularEngines = 0;
@@ -121,10 +120,6 @@ void ExecutionEnvironment::calculateMaxOsContextCount() {
}
MemoryManager::maxOsContextCount += (numSecondaryContexts + osContextCount) * subDevicesCount + numRootContexts;
if (ccsCount > 1 && debugManager.flags.EngineInstancedSubDevices.get()) {
MemoryManager::maxOsContextCount += ccsCount * subDevicesCount;
}
}
}
@@ -270,7 +265,7 @@ void ExecutionEnvironment::parseAffinityMask() {
continue;
}
// FlatHierarchy not supported with AllowSingleTileEngineInstancedSubDevices
// FlatHierarchy
// so ignore X.Y
if (subEntries.size() > 1) {
continue;
@@ -293,19 +288,7 @@ void ExecutionEnvironment::parseAffinityMask() {
if (subEntries.size() > 1) {
uint32_t subDeviceIndex = StringHelpers::toUint32t(subEntries[1]);
bool enableSecondLevelEngineInstanced = ((subDevicesCount == 1) &&
(hwInfo->gtSystemInfo.CCSInfo.NumberOfCCSEnabled > 1) &&
debugManager.flags.AllowSingleTileEngineInstancedSubDevices.get());
if (enableSecondLevelEngineInstanced) {
UNRECOVERABLE_IF(subEntries.size() != 2);
if (subDeviceIndex < hwInfo->gtSystemInfo.CCSInfo.NumberOfCCSEnabled) {
// Store the Physical Hierarchy for this SubDevice mapped to the Device Index passed to the user.
mapOfSubDeviceIndices[rootDeviceIndex] = std::make_tuple(rootDeviceIndex, subDeviceIndex, subDevicesCount);
affinityMaskHelper[rootDeviceIndex].enableEngineInstancedSubDevice(0, subDeviceIndex); // Mask: X.Y
}
} else if (subDeviceIndex < subDevicesCount) {
if (subDeviceIndex < subDevicesCount) {
if (subEntries.size() == 2) {
// Store the Physical Hierarchy for this SubDevice mapped to the Device Index passed to the user.
mapOfSubDeviceIndices[rootDeviceIndex] = std::make_tuple(rootDeviceIndex, subDeviceIndex, subDevicesCount);

7
shared/source/helpers/engine_node_helper.h
View File

@@ -1,5 +1,5 @@
/*
* Copyright (C) 2019-2023 Intel Corporation
* Copyright (C) 2019-2024 Intel Corporation
*
* SPDX-License-Identifier: MIT
*
@@ -34,14 +34,13 @@ using EngineTypeUsage = std::pair<aub_stream::EngineType, EngineUsage>;
struct EngineDescriptor {
EngineDescriptor() = delete;
constexpr EngineDescriptor(EngineTypeUsage engineTypeUsage, DeviceBitfield deviceBitfield, PreemptionMode preemptionMode, bool isRootDevice, bool isEngineInstanced)
: engineTypeUsage(engineTypeUsage), deviceBitfield(deviceBitfield), preemptionMode(preemptionMode), isRootDevice(isRootDevice), isEngineInstanced(isEngineInstanced) {}
constexpr EngineDescriptor(EngineTypeUsage engineTypeUsage, DeviceBitfield deviceBitfield, PreemptionMode preemptionMode, bool isRootDevice)
: engineTypeUsage(engineTypeUsage), deviceBitfield(deviceBitfield), preemptionMode(preemptionMode), isRootDevice(isRootDevice) {}
EngineTypeUsage engineTypeUsage;
DeviceBitfield deviceBitfield;
PreemptionMode preemptionMode;
bool isRootDevice;
bool isEngineInstanced;
};
namespace EngineHelpers {

4
shared/source/helpers/gfx_core_helper.h
View File

@@ -124,7 +124,7 @@ class GfxCoreHelper {
virtual bool isRcsAvailable(const HardwareInfo &hwInfo) const = 0;
virtual bool isCooperativeDispatchSupported(const EngineGroupType engineGroupType, const RootDeviceEnvironment &rootDeviceEnvironment) const = 0;
virtual uint32_t adjustMaxWorkGroupCount(uint32_t maxWorkGroupCount, const EngineGroupType engineGroupType,
const RootDeviceEnvironment &rootDeviceEnvironment, bool isEngineInstanced) const = 0;
const RootDeviceEnvironment &rootDeviceEnvironment) const = 0;
virtual uint32_t adjustMaxWorkGroupSize(const uint32_t grfCount, const uint32_t simd, bool isHwLocalGeneration, const uint32_t defaultMaxGroupSize, const RootDeviceEnvironment &rootDeviceEnvironment) const = 0;
virtual size_t getMaxFillPaternSizeForCopyEngine() const = 0;
virtual size_t getSipKernelMaxDbgSurfaceSize(const HardwareInfo &hwInfo) const = 0;
@@ -347,7 +347,7 @@ class GfxCoreHelperHw : public GfxCoreHelper {
bool isCooperativeDispatchSupported(const EngineGroupType engineGroupType, const RootDeviceEnvironment &rootDeviceEnvironment) const override;
uint32_t adjustMaxWorkGroupCount(uint32_t maxWorkGroupCount, const EngineGroupType engineGroupType,
const RootDeviceEnvironment &rootDeviceEnvironment, bool isEngineInstanced) const override;
const RootDeviceEnvironment &rootDeviceEnvironment) const override;
uint32_t adjustMaxWorkGroupSize(const uint32_t grfCount, const uint32_t simd, bool isHwLocalGeneration, const uint32_t defaultMaxGroupSize, const RootDeviceEnvironment &rootDeviceEnvironment) const override;
size_t getMaxFillPaternSizeForCopyEngine() const override;

4
shared/source/helpers/gfx_core_helper_bdw_to_dg2.inl
View File

@@ -1,5 +1,5 @@
/*
* Copyright (C) 2021-2023 Intel Corporation
* Copyright (C) 2021-2024 Intel Corporation
*
* SPDX-License-Identifier: MIT
*
@@ -30,7 +30,7 @@ bool GfxCoreHelperHw<GfxFamily>::isCooperativeDispatchSupported(const EngineGrou
template <typename GfxFamily>
uint32_t GfxCoreHelperHw<GfxFamily>::adjustMaxWorkGroupCount(uint32_t maxWorkGroupCount, const EngineGroupType engineGroupType,
const RootDeviceEnvironment &rootDeviceEnvironment, bool isEngineInstanced) const {
const RootDeviceEnvironment &rootDeviceEnvironment) const {
return maxWorkGroupCount;
}

5
shared/source/helpers/gfx_core_helper_pvc_and_later.inl
View File

@@ -50,7 +50,7 @@ bool GfxCoreHelperHw<Family>::isCooperativeDispatchSupported(const EngineGroupTy
template <typename Family>
uint32_t GfxCoreHelperHw<Family>::adjustMaxWorkGroupCount(uint32_t maxWorkGroupCount, const EngineGroupType engineGroupType,
const RootDeviceEnvironment &rootDeviceEnvironment, bool isEngineInstanced) const {
const RootDeviceEnvironment &rootDeviceEnvironment) const {
if ((debugManager.flags.ForceTheoreticalMaxWorkGroupCount.get()) ||
(debugManager.flags.OverrideMaxWorkGroupCount.get() != -1)) {
return maxWorkGroupCount;
@@ -61,8 +61,7 @@ uint32_t GfxCoreHelperHw<Family>::adjustMaxWorkGroupCount(uint32_t maxWorkGroupC
auto &productHelper = rootDeviceEnvironment.getHelper<ProductHelper>();
auto &hwInfo = *rootDeviceEnvironment.getHardwareInfo();
bool requiresLimitation = productHelper.isCooperativeEngineSupported(hwInfo) &&
(engineGroupType != EngineGroupType::cooperativeCompute) &&
(!isEngineInstanced);
(engineGroupType != EngineGroupType::cooperativeCompute);
auto ccsCount = hwInfo.gtSystemInfo.CCSInfo.NumberOfCCSEnabled;
auto numberOfpartsInTileForConcurrentKernels = productHelper.getNumberOfPartsInTileForConcurrentKernel(ccsCount);

8
shared/source/helpers/kernel_helpers.cpp
View File

@@ -21,13 +21,13 @@
namespace NEO {
uint32_t KernelHelper::getMaxWorkGroupCount(const RootDeviceEnvironment &rootDeviceEnvironment, const KernelDescriptor &kernelDescriptor, uint32_t numSubDevices,
uint32_t usedSlmSize, uint32_t workDim, const size_t *localWorkSize, EngineGroupType engineGroupType, bool isEngineInstanced) {
uint32_t usedSlmSize, uint32_t workDim, const size_t *localWorkSize, EngineGroupType engineGroupType) {
return KernelHelper::getMaxWorkGroupCount(rootDeviceEnvironment, kernelDescriptor.kernelAttributes.numGrfRequired, kernelDescriptor.kernelAttributes.simdSize, kernelDescriptor.kernelAttributes.barrierCount,
numSubDevices, usedSlmSize, workDim, localWorkSize, engineGroupType, isEngineInstanced);
numSubDevices, usedSlmSize, workDim, localWorkSize, engineGroupType);
}
uint32_t KernelHelper::getMaxWorkGroupCount(const RootDeviceEnvironment &rootDeviceEnvironment, uint16_t numGrfRequired, uint8_t simdSize, uint8_t barrierCount,
uint32_t numSubDevices, uint32_t usedSlmSize, uint32_t workDim, const size_t *localWorkSize, EngineGroupType engineGroupType, bool isEngineInstanced) {
uint32_t numSubDevices, uint32_t usedSlmSize, uint32_t workDim, const size_t *localWorkSize, EngineGroupType engineGroupType) {
if (debugManager.flags.OverrideMaxWorkGroupCount.get() != -1) {
return static_cast<uint32_t>(debugManager.flags.OverrideMaxWorkGroupCount.get());
}
@@ -65,7 +65,7 @@ uint32_t KernelHelper::getMaxWorkGroupCount(const RootDeviceEnvironment &rootDev
maxWorkGroupsCount = std::min(maxWorkGroupsCount, maxWorkGroupsCountDueToSlm);
}
maxWorkGroupsCount = helper.adjustMaxWorkGroupCount(maxWorkGroupsCount, engineGroupType, rootDeviceEnvironment, isEngineInstanced);
maxWorkGroupsCount = helper.adjustMaxWorkGroupCount(maxWorkGroupsCount, engineGroupType, rootDeviceEnvironment);
if (!helper.singleTileExecImplicitScalingRequired(true)) {
maxWorkGroupsCount *= numSubDevices;

4
shared/source/helpers/kernel_helpers.h
View File

@@ -24,9 +24,9 @@ struct KernelHelper {
invalidKernel = 2
};
static uint32_t getMaxWorkGroupCount(const RootDeviceEnvironment &rootDeviceEnvironment, const KernelDescriptor &kernelDescriptor,
uint32_t numSubDevices, uint32_t usedSlmSize, uint32_t workDim, const size_t *localWorkSize, EngineGroupType engineGroupType, bool isEngineInstanced);
uint32_t numSubDevices, uint32_t usedSlmSize, uint32_t workDim, const size_t *localWorkSize, EngineGroupType engineGroupType);
static uint32_t getMaxWorkGroupCount(const RootDeviceEnvironment &rootDeviceEnvironment, uint16_t numGrfRequired, uint8_t simdSize, uint8_t barrierCount,
uint32_t numSubDevices, uint32_t usedSlmSize, uint32_t workDim, const size_t *localWorkSize, EngineGroupType engineGroupType, bool isEngineInstanced);
uint32_t numSubDevices, uint32_t usedSlmSize, uint32_t workDim, const size_t *localWorkSize, EngineGroupType engineGroupType);
static inline uint64_t getPrivateSurfaceSize(uint64_t perHwThreadPrivateMemorySize, uint32_t computeUnitsUsedForScratch) {
return perHwThreadPrivateMemorySize * computeUnitsUsedForScratch;
}

4
shared/source/os_interface/linux/drm_neo.cpp
View File

@@ -1166,7 +1166,7 @@ void Drm::configureGpuFaultCheckThreshold() {
}
}
unsigned int Drm::bindDrmContext(uint32_t drmContextId, uint32_t deviceIndex, aub_stream::EngineType engineType, bool engineInstancedDevice) {
unsigned int Drm::bindDrmContext(uint32_t drmContextId, uint32_t deviceIndex, aub_stream::EngineType engineType) {
auto engineInfo = this->engineInfo.get();
auto retVal = static_cast<unsigned int>(ioctlHelper->getDrmParamValue(DrmEngineMapper::engineNodeMap(engineType)));
@@ -1179,7 +1179,7 @@ unsigned int Drm::bindDrmContext(uint32_t drmContextId, uint32_t deviceIndex, au
return retVal;
}
bool useVirtualEnginesForCcs = !engineInstancedDevice;
bool useVirtualEnginesForCcs = true;
if (debugManager.flags.UseDrmVirtualEnginesForCcs.get() != -1) {
useVirtualEnginesForCcs = !!debugManager.flags.UseDrmVirtualEnginesForCcs.get();
}

2
shared/source/os_interface/linux/drm_neo.h
View File

@@ -106,7 +106,7 @@ class Drm : public DriverModel {
int queryVmId(uint32_t drmContextId, uint32_t &vmId);
void setLowPriorityContextParam(uint32_t drmContextId);
MOCKABLE_VIRTUAL unsigned int bindDrmContext(uint32_t drmContextId, uint32_t deviceIndex, aub_stream::EngineType engineType, bool engineInstancedDevice);
MOCKABLE_VIRTUAL unsigned int bindDrmContext(uint32_t drmContextId, uint32_t deviceIndex, aub_stream::EngineType engineType);
MOCKABLE_VIRTUAL int getErrno();
bool setQueueSliceCount(uint64_t sliceCount);

2
shared/source/os_interface/linux/ioctl_helper_i915.cpp
View File

@@ -442,7 +442,7 @@ int IoctlHelperI915::createDrmContext(Drm &drm, OsContextLinux &osContext, uint3
if (drm.isPreemptionSupported() && osContext.isLowPriority()) {
drm.setLowPriorityContextParam(drmContextId);
}
auto engineFlag = drm.bindDrmContext(drmContextId, deviceIndex, osContext.getEngineType(), osContext.isEngineInstanced());
auto engineFlag = drm.bindDrmContext(drmContextId, deviceIndex, osContext.getEngineType());
osContext.setEngineFlag(engineFlag);
return drmContextId;
}

2
shared/source/os_interface/linux/xe/ioctl_helper_xe.cpp
View File

@@ -1239,7 +1239,7 @@ int IoctlHelperXe::createDrmContext(Drm &drm, OsContextLinux &osContext, uint32_
uint32_t drmContextId = 0;
xeLog("createDrmContext VM=0x%x\n", drmVmId);
drm.bindDrmContext(drmContextId, deviceIndex, osContext.getEngineType(), osContext.isEngineInstanced());
drm.bindDrmContext(drmContextId, deviceIndex, osContext.getEngineType());
UNRECOVERABLE_IF(contextParamEngine.empty());

3
shared/source/os_interface/os_context.cpp
View File

@@ -22,8 +22,7 @@ OsContext::OsContext(uint32_t rootDeviceIndex, uint32_t contextId, const EngineD
numSupportedDevices(static_cast<uint32_t>(engineDescriptor.deviceBitfield.count())),
engineType(engineDescriptor.engineTypeUsage.first),
engineUsage(engineDescriptor.engineTypeUsage.second),
rootDevice(engineDescriptor.isRootDevice),
engineInstancedDevice(engineDescriptor.isEngineInstanced) {}
rootDevice(engineDescriptor.isRootDevice) {}
bool OsContext::isImmediateContextInitializationEnabled(bool isDefaultEngine) const {
if (debugManager.flags.DeferOsContextInitialization.get() == 0) {

2
shared/source/os_interface/os_context.h
View File

@@ -42,7 +42,6 @@ class OsContext : public ReferenceTrackedObject<OsContext> {
bool isInternalEngine() const { return engineUsage == EngineUsage::internal; }
bool isCooperativeEngine() const { return engineUsage == EngineUsage::cooperative; }
bool isRootDevice() const { return rootDevice; }
bool isEngineInstanced() const { return engineInstancedDevice; }
virtual bool isDirectSubmissionSupported() const { return false; }
bool isDefaultContext() const { return defaultContext; }
void setDefaultContext(bool value) { defaultContext = value; }
@@ -119,7 +118,6 @@ class OsContext : public ReferenceTrackedObject<OsContext> {
std::once_flag contextInitializedFlag = {};
bool contextInitialized = false;
bool debuggableContext = false;
bool engineInstancedDevice = false;
uint8_t powerHintValue = 0;
bool isContextGroup = false;

26
shared/test/common/helpers/engine_descriptor_helper.h
View File

@@ -1,5 +1,5 @@
/*
* Copyright (C) 2021-2023 Intel Corporation
* Copyright (C) 2021-2024 Intel Corporation
*
* SPDX-License-Identifier: MIT
*
@@ -15,63 +15,55 @@ constexpr EngineDescriptor getDefaultDescriptor() {
return {{aub_stream::EngineType::ENGINE_RCS, EngineUsage::regular},
1 /*deviceBitfield*/,
PreemptionMode::Disabled,
false /* isRootDevice*/,
false /* isEngineInstanced*/};
false /* isRootDevice*/};
}
constexpr EngineDescriptor getDefaultDescriptor(EngineTypeUsage engineTypeUsage) {
return {engineTypeUsage,
1 /*deviceBitfield*/,
PreemptionMode::Disabled,
false /* isRootDevice*/,
false /* isEngineInstanced*/};
false /* isRootDevice*/};
}
constexpr EngineDescriptor getDefaultDescriptor(EngineTypeUsage engineTypeUsage, bool isRootDevice) {
return {engineTypeUsage,
1 /*deviceBitfield*/,
PreemptionMode::Disabled,
isRootDevice,
false /* isEngineInstanced*/};
isRootDevice};
}
constexpr EngineDescriptor getDefaultDescriptor(EngineTypeUsage engineTypeUsage, PreemptionMode preemptionMode) {
return {engineTypeUsage,
1 /*deviceBitfield*/,
preemptionMode,
false /* isRootDevice*/,
false /* isEngineInstanced*/};
false /* isRootDevice*/};
}
inline EngineDescriptor getDefaultDescriptor(DeviceBitfield deviceBitfield) {
return {{aub_stream::EngineType::ENGINE_RCS, EngineUsage::regular},
deviceBitfield,
PreemptionMode::Disabled,
false /* isRootDevice*/,
false /* isEngineInstanced*/};
false /* isRootDevice*/};
}
inline EngineDescriptor getDefaultDescriptor(EngineTypeUsage engineTypeUsage, DeviceBitfield deviceBitfield) {
return {engineTypeUsage,
deviceBitfield,
PreemptionMode::Disabled,
false /* isRootDevice*/,
false /* isEngineInstanced*/};
false /* isRootDevice*/};
}
inline EngineDescriptor getDefaultDescriptor(EngineTypeUsage engineTypeUsage, PreemptionMode preemptionMode, DeviceBitfield deviceBitfield) {
return {engineTypeUsage,
deviceBitfield,
preemptionMode,
false /* isRootDevice*/,
false /* isEngineInstanced*/};
false /* isRootDevice*/};
}
inline EngineDescriptor getDefaultDescriptor(EngineTypeUsage engineTypeUsage, PreemptionMode preemptionMode, DeviceBitfield deviceBitfield, bool isRootDevice) {
return {engineTypeUsage,
deviceBitfield,
preemptionMode,
isRootDevice,
false /* isEngineInstanced*/};
isRootDevice};
}
} // namespace EngineDescriptorHelper

2
shared/test/common/mocks/mock_device.cpp
View File

@@ -35,7 +35,7 @@ MockDevice::MockDevice()
commandStreamReceivers.resize(1);
commandStreamReceivers[0].reset(commandStreamReceiver);
EngineDescriptor engineDescriptor = {EngineTypeUsage{aub_stream::ENGINE_CCS, EngineUsage::regular}, this->getDeviceBitfield(), PreemptionMode::Disabled, true, false};
EngineDescriptor engineDescriptor = {EngineTypeUsage{aub_stream::ENGINE_CCS, EngineUsage::regular}, this->getDeviceBitfield(), PreemptionMode::Disabled, true};
OsContext *osContext = getMemoryManager()->createAndRegisterOsContext(commandStreamReceiver, engineDescriptor);
commandStreamReceiver->setupContext(*osContext);

8
shared/test/common/mocks/mock_device.h
View File

@@ -33,8 +33,6 @@ extern CommandStreamReceiver *createCommandStream(ExecutionEnvironment &executio
struct MockSubDevice : public SubDevice {
using Device::allEngines;
using Device::createEngines;
using Device::engineInstancedType;
using SubDevice::engineInstanced;
using SubDevice::getDeviceBitfield;
using SubDevice::getGlobalMemorySize;
using SubDevice::SubDevice;
@@ -63,8 +61,6 @@ class MockDevice : public RootDevice {
using Device::createSubDevices;
using Device::deviceBitfield;
using Device::deviceInfo;
using Device::engineInstanced;
using Device::engineInstancedType;
using Device::executionEnvironment;
using Device::generateUuidFromPciBusInfo;
using Device::getGlobalMemorySize;
@@ -153,10 +149,6 @@ class MockDevice : public RootDevice {
return Device::create<MockSubDevice>(executionEnvironment, subDeviceIndex, *this);
}
SubDevice *createEngineInstancedSubDevice(uint32_t subDeviceIndex, aub_stream::EngineType engineType) override {
return Device::create<MockSubDevice>(executionEnvironment, subDeviceIndex, *this, engineType);
}
std::unique_ptr<CommandStreamReceiver> createCommandStreamReceiver() const override {
return std::unique_ptr<CommandStreamReceiver>(createCommandStreamReceiverFunc(*executionEnvironment, getRootDeviceIndex(), getDeviceBitfield()));
}

2
shared/test/common/test_files/igdrcl.config
View File

@@ -294,7 +294,6 @@ DeferOsContextInitialization = -1
DebuggerForceSbaTrackingMode = -1
ExperimentalEnableCustomLocalMemoryAlignment = 0
AlignLocalMemoryVaTo2MB = -1
EngineInstancedSubDevices = 0
OverrideTimestampPacketSize = -1
ComputeOverdispatchDisable = -1
CFEWeightedDispatchModeDisable = -1
@@ -393,7 +392,6 @@ OverrideBufferSuitableForRenderCompression = -1
AllowPatchingVfeStateInCommandLists = 0
PrintMemoryRegionSizes = 0
OverrideDrmRegion = -1
AllowSingleTileEngineInstancedSubDevices = 0
BinaryCacheTrace = false
OverrideL1CacheControlInSurfaceState = -1
OverrideL1CacheControlInSurfaceStateForScratchSpace = -1

77
shared/test/unit_test/command_stream/stream_properties_tests_common.cpp
View File

@@ -85,7 +85,7 @@ TEST(StreamPropertiesTests, whenSettingCooperativeKernelPropertiesThenCorrectVal
for (auto isCooperativeKernel : ::testing::Bool()) {
for (auto disableOverdispatch : ::testing::Bool()) {
for (auto disableEUFusion : ::testing::Bool()) {
properties.frontEndState.setPropertiesAll(isCooperativeKernel, disableEUFusion, disableOverdispatch, false);
properties.frontEndState.setPropertiesAll(isCooperativeKernel, disableEUFusion, disableOverdispatch);
if (frontEndPropertiesSupport.computeDispatchAllWalker) {
EXPECT_EQ(isCooperativeKernel, properties.frontEndState.computeDispatchAllWalkerEnable.value);
} else {
@@ -102,7 +102,7 @@ TEST(StreamPropertiesTests, whenSettingCooperativeKernelPropertiesThenCorrectVal
EXPECT_EQ(-1, properties.frontEndState.disableOverdispatch.value);
}
if (frontEndPropertiesSupport.singleSliceDispatchCcsMode) {
EXPECT_EQ(false, properties.frontEndState.singleSliceDispatchCcsMode.value);
EXPECT_EQ(-1, properties.frontEndState.singleSliceDispatchCcsMode.value);
} else {
EXPECT_EQ(-1, properties.frontEndState.singleSliceDispatchCcsMode.value);
}
@@ -512,88 +512,75 @@ TEST(StreamPropertiesTests, givenSingleDispatchCcsFrontEndPropertyWhenSettingPro
feProperties.initSupport(*mockExecutionEnvironment.rootDeviceEnvironments[0]);
EXPECT_TRUE(feProperties.propertiesSupportLoaded);
int32_t engineInstancedDevice = 1;
feProperties.setPropertySingleSliceDispatchCcsMode(engineInstancedDevice);
if (fePropertiesSupport.singleSliceDispatchCcsMode) {
EXPECT_TRUE(feProperties.singleSliceDispatchCcsMode.isDirty);
EXPECT_EQ(engineInstancedDevice, feProperties.singleSliceDispatchCcsMode.value);
EXPECT_FALSE(feProperties.singleSliceDispatchCcsMode.isDirty);
} else {
EXPECT_FALSE(feProperties.singleSliceDispatchCcsMode.isDirty);
EXPECT_EQ(-1, feProperties.singleSliceDispatchCcsMode.value);
}
feProperties.frontEndPropertiesSupport.singleSliceDispatchCcsMode = true;
engineInstancedDevice = 2;
feProperties.setPropertySingleSliceDispatchCcsMode(engineInstancedDevice);
EXPECT_TRUE(feProperties.singleSliceDispatchCcsMode.isDirty);
EXPECT_EQ(engineInstancedDevice, feProperties.singleSliceDispatchCcsMode.value);
EXPECT_FALSE(feProperties.singleSliceDispatchCcsMode.isDirty);
}
TEST(StreamPropertiesTests, givenDisableOverdispatchEngineInstancedFrontEndPropertyWhenSettingPropertyAndCheckIfSupportedThenExpectCorrectState) {
TEST(StreamPropertiesTests, givenDisableOverdispatchFrontEndPropertyWhenSettingPropertyAndCheckIfSupportedThenExpectCorrectState) {
bool clearDirtyState = false;
MockFrontEndProperties feProperties{};
feProperties.propertiesSupportLoaded = true;
feProperties.frontEndPropertiesSupport.disableOverdispatch = false;
feProperties.frontEndPropertiesSupport.singleSliceDispatchCcsMode = false;
int32_t engineInstancedDevice = 0;
bool disableOverdispatch = false;
feProperties.setPropertiesDisableOverdispatchEngineInstanced(disableOverdispatch, engineInstancedDevice, clearDirtyState);
feProperties.setPropertiesDisableOverdispatch(disableOverdispatch, clearDirtyState);
EXPECT_FALSE(feProperties.isDirty());
EXPECT_EQ(-1, feProperties.disableOverdispatch.value);
EXPECT_EQ(-1, feProperties.singleSliceDispatchCcsMode.value);
feProperties.frontEndPropertiesSupport.disableOverdispatch = true;
feProperties.frontEndPropertiesSupport.singleSliceDispatchCcsMode = true;
engineInstancedDevice = -1;
feProperties.setPropertiesDisableOverdispatchEngineInstanced(disableOverdispatch, engineInstancedDevice, clearDirtyState);
feProperties.setPropertiesDisableOverdispatch(disableOverdispatch, clearDirtyState);
EXPECT_TRUE(feProperties.isDirty());
EXPECT_EQ(0, feProperties.disableOverdispatch.value);
EXPECT_EQ(-1, feProperties.singleSliceDispatchCcsMode.value);
engineInstancedDevice = 0;
feProperties.setPropertiesAll(false, false, disableOverdispatch, engineInstancedDevice);
EXPECT_TRUE(feProperties.isDirty());
EXPECT_EQ(0, feProperties.disableOverdispatch.value);
EXPECT_EQ(0, feProperties.singleSliceDispatchCcsMode.value);
feProperties.setPropertiesDisableOverdispatchEngineInstanced(disableOverdispatch, engineInstancedDevice, clearDirtyState);
feProperties.setPropertiesAll(false, false, disableOverdispatch);
EXPECT_FALSE(feProperties.isDirty());
EXPECT_EQ(0, feProperties.disableOverdispatch.value);
EXPECT_EQ(0, feProperties.singleSliceDispatchCcsMode.value);
EXPECT_EQ(-1, feProperties.singleSliceDispatchCcsMode.value);
engineInstancedDevice = 1;
feProperties.setPropertiesDisableOverdispatchEngineInstanced(disableOverdispatch, engineInstancedDevice, clearDirtyState);
EXPECT_TRUE(feProperties.isDirty());
EXPECT_EQ(0, feProperties.disableOverdispatch.value);
EXPECT_EQ(1, feProperties.singleSliceDispatchCcsMode.value);
feProperties.setPropertiesDisableOverdispatchEngineInstanced(disableOverdispatch, engineInstancedDevice, clearDirtyState);
feProperties.setPropertiesDisableOverdispatch(disableOverdispatch, clearDirtyState);
EXPECT_FALSE(feProperties.isDirty());
EXPECT_EQ(0, feProperties.disableOverdispatch.value);
EXPECT_EQ(1, feProperties.singleSliceDispatchCcsMode.value);
EXPECT_EQ(-1, feProperties.singleSliceDispatchCcsMode.value);
feProperties.setPropertiesDisableOverdispatch(disableOverdispatch, clearDirtyState);
EXPECT_FALSE(feProperties.isDirty());
EXPECT_EQ(0, feProperties.disableOverdispatch.value);
EXPECT_EQ(-1, feProperties.singleSliceDispatchCcsMode.value);
feProperties.setPropertiesDisableOverdispatch(disableOverdispatch, clearDirtyState);
EXPECT_FALSE(feProperties.isDirty());
EXPECT_EQ(0, feProperties.disableOverdispatch.value);
EXPECT_EQ(-1, feProperties.singleSliceDispatchCcsMode.value);
disableOverdispatch = true;
feProperties.setPropertiesDisableOverdispatchEngineInstanced(disableOverdispatch, engineInstancedDevice, clearDirtyState);
feProperties.setPropertiesDisableOverdispatch(disableOverdispatch, clearDirtyState);
EXPECT_TRUE(feProperties.isDirty());
EXPECT_EQ(1, feProperties.disableOverdispatch.value);
EXPECT_EQ(1, feProperties.singleSliceDispatchCcsMode.value);
EXPECT_EQ(-1, feProperties.singleSliceDispatchCcsMode.value);
feProperties.setPropertiesDisableOverdispatchEngineInstanced(disableOverdispatch, engineInstancedDevice, clearDirtyState);
feProperties.setPropertiesDisableOverdispatch(disableOverdispatch, clearDirtyState);
EXPECT_FALSE(feProperties.isDirty());
EXPECT_EQ(1, feProperties.disableOverdispatch.value);
EXPECT_EQ(1, feProperties.singleSliceDispatchCcsMode.value);
EXPECT_EQ(-1, feProperties.singleSliceDispatchCcsMode.value);
clearDirtyState = true;
disableOverdispatch = false;
engineInstancedDevice = 0;
feProperties.setPropertiesDisableOverdispatchEngineInstanced(disableOverdispatch, engineInstancedDevice, clearDirtyState);
feProperties.setPropertiesDisableOverdispatch(disableOverdispatch, clearDirtyState);
EXPECT_FALSE(feProperties.isDirty());
EXPECT_EQ(0, feProperties.disableOverdispatch.value);
EXPECT_EQ(0, feProperties.singleSliceDispatchCcsMode.value);
EXPECT_EQ(-1, feProperties.singleSliceDispatchCcsMode.value);
}
TEST(StreamPropertiesTests, givenComputeDispatchAllWalkerEnableAndDisableEuFusionFrontEndPropertiesWhenSettingPropertiesAndCheckIfSupportedThenExpectCorrectState) {
@@ -611,7 +598,7 @@ TEST(StreamPropertiesTests, givenComputeDispatchAllWalkerEnableAndDisableEuFusio
feProperties.frontEndPropertiesSupport.disableEuFusion = true;
feProperties.frontEndPropertiesSupport.computeDispatchAllWalker = true;
feProperties.setPropertiesAll(isCooperativeKernel, disableEuFusion, false, -1);
feProperties.setPropertiesAll(isCooperativeKernel, disableEuFusion, false);
EXPECT_TRUE(feProperties.isDirty());
EXPECT_EQ(0, feProperties.disableEUFusion.value);
EXPECT_EQ(0, feProperties.computeDispatchAllWalkerEnable.value);
@@ -685,13 +672,12 @@ TEST(StreamPropertiesTests, givenSetAllFrontEndPropertiesWhenResettingStateThenR
bool isCooperativeKernel = false;
bool disableEuFusion = true;
bool disableOverdispatch = true;
int32_t engineInstancedDevice = 3;
feProperties.setPropertiesAll(isCooperativeKernel, disableEuFusion, disableOverdispatch, engineInstancedDevice);
feProperties.setPropertiesAll(isCooperativeKernel, disableEuFusion, disableOverdispatch);
EXPECT_TRUE(feProperties.isDirty());
EXPECT_EQ(0, feProperties.computeDispatchAllWalkerEnable.value);
EXPECT_EQ(1, feProperties.disableEUFusion.value);
EXPECT_EQ(1, feProperties.disableOverdispatch.value);
EXPECT_EQ(3, feProperties.singleSliceDispatchCcsMode.value);
EXPECT_EQ(-1, feProperties.singleSliceDispatchCcsMode.value);
feProperties.resetState();
EXPECT_FALSE(feProperties.isDirty());
@@ -1446,8 +1432,7 @@ TEST(StreamPropertiesTests, givenAllStreamPropertiesSetWhenAllStreamPropertiesRe
bool isCooperativeKernel = false;
bool disableEuFusion = true;
bool disableOverdispatch = true;
int32_t engineInstancedDevice = 3;
globalStreamProperties.frontEndState.setPropertiesAll(isCooperativeKernel, disableEuFusion, disableOverdispatch, engineInstancedDevice);
globalStreamProperties.frontEndState.setPropertiesAll(isCooperativeKernel, disableEuFusion, disableOverdispatch);
bool modeSelected = false;
bool mediaSamplerDopClockGate = false;

2
shared/test/unit_test/execution_environment/execution_environment_tests.cpp
View File

@@ -463,7 +463,6 @@ TEST(ExecutionEnvironment, whenCalculateMaxOsContexCountThenGlobalVariableHasPro
uint32_t expectedOsContextCountForCcs = 0u;
{
debugManager.flags.EngineInstancedSubDevices.set(false);
MockExecutionEnvironment executionEnvironment(nullptr, true, numRootDevices);
for (const auto &rootDeviceEnvironment : executionEnvironment.rootDeviceEnvironments) {
@@ -485,7 +484,6 @@ TEST(ExecutionEnvironment, whenCalculateMaxOsContexCountThenGlobalVariableHasPro
}
{
debugManager.flags.EngineInstancedSubDevices.set(true);
MockExecutionEnvironment executionEnvironment(nullptr, true, numRootDevices);
EXPECT_EQ(expectedOsContextCount + expectedOsContextCountForCcs, MemoryManager::maxOsContextCount);

20
shared/test/unit_test/helpers/gfx_core_helper_tests.cpp
View File

@@ -1611,22 +1611,22 @@ HWTEST_F(GfxCoreHelperTest, GivenCooperativeEngineSupportedAndNotUsedWhenAdjustM
((engineGroupType == EngineGroupType::compute) && isRcsEnabled);
bool applyLimitation = engineGroupType != EngineGroupType::cooperativeCompute;
if (disallowDispatch) {
EXPECT_EQ(1u, gfxCoreHelper.adjustMaxWorkGroupCount(4u, engineGroupType, rootDeviceEnvironment, false));
EXPECT_EQ(1u, gfxCoreHelper.adjustMaxWorkGroupCount(1024u, engineGroupType, rootDeviceEnvironment, false));
EXPECT_EQ(1u, gfxCoreHelper.adjustMaxWorkGroupCount(4u, engineGroupType, rootDeviceEnvironment));
EXPECT_EQ(1u, gfxCoreHelper.adjustMaxWorkGroupCount(1024u, engineGroupType, rootDeviceEnvironment));
} else if (applyLimitation) {
hwInfo.gtSystemInfo.CCSInfo.NumberOfCCSEnabled = 4;
EXPECT_EQ(1u, gfxCoreHelper.adjustMaxWorkGroupCount(4u, engineGroupType, rootDeviceEnvironment, false));
EXPECT_EQ(256u, gfxCoreHelper.adjustMaxWorkGroupCount(1024u, engineGroupType, rootDeviceEnvironment, false));
EXPECT_EQ(1u, gfxCoreHelper.adjustMaxWorkGroupCount(4u, engineGroupType, rootDeviceEnvironment));
EXPECT_EQ(256u, gfxCoreHelper.adjustMaxWorkGroupCount(1024u, engineGroupType, rootDeviceEnvironment));
hwInfo.gtSystemInfo.CCSInfo.NumberOfCCSEnabled = 2;
EXPECT_EQ(2u, gfxCoreHelper.adjustMaxWorkGroupCount(4u, engineGroupType, rootDeviceEnvironment, false));
EXPECT_EQ(512u, gfxCoreHelper.adjustMaxWorkGroupCount(1024u, engineGroupType, rootDeviceEnvironment, false));
EXPECT_EQ(2u, gfxCoreHelper.adjustMaxWorkGroupCount(4u, engineGroupType, rootDeviceEnvironment));
EXPECT_EQ(512u, gfxCoreHelper.adjustMaxWorkGroupCount(1024u, engineGroupType, rootDeviceEnvironment));
} else {
EXPECT_EQ(4u, gfxCoreHelper.adjustMaxWorkGroupCount(4u, engineGroupType, rootDeviceEnvironment, false));
EXPECT_EQ(1024u, gfxCoreHelper.adjustMaxWorkGroupCount(1024u, engineGroupType, rootDeviceEnvironment, false));
EXPECT_EQ(4u, gfxCoreHelper.adjustMaxWorkGroupCount(4u, engineGroupType, rootDeviceEnvironment));
EXPECT_EQ(1024u, gfxCoreHelper.adjustMaxWorkGroupCount(1024u, engineGroupType, rootDeviceEnvironment));
}
} else {
EXPECT_EQ(4u, gfxCoreHelper.adjustMaxWorkGroupCount(4u, engineGroupType, rootDeviceEnvironment, false));
EXPECT_EQ(1024u, gfxCoreHelper.adjustMaxWorkGroupCount(1024u, engineGroupType, rootDeviceEnvironment, false));
EXPECT_EQ(4u, gfxCoreHelper.adjustMaxWorkGroupCount(4u, engineGroupType, rootDeviceEnvironment));
EXPECT_EQ(1024u, gfxCoreHelper.adjustMaxWorkGroupCount(1024u, engineGroupType, rootDeviceEnvironment));
}
}
}

6
shared/test/unit_test/helpers/kernel_helpers_tests.cpp
View File

@@ -46,7 +46,7 @@ struct KernelHelperMaxWorkGroupsTests : ::testing::Test {
hwInfo->gtSystemInfo.DualSubSliceCount = dssCount;
hwInfo->capabilityTable.slmSize = (availableSlm / MemoryConstants::kiloByte) / dssCount;
return KernelHelper::getMaxWorkGroupCount(*rootDeviceEnvironment, descriptor, numSubdevices, usedSlm, workDim, lws, engineType, false);
return KernelHelper::getMaxWorkGroupCount(*rootDeviceEnvironment, descriptor, numSubdevices, usedSlm, workDim, lws, engineType);
}
std::unique_ptr<MockExecutionEnvironment> executionEnvironment;
@@ -59,7 +59,7 @@ TEST_F(KernelHelperMaxWorkGroupsTests, GivenNoBarriersOrSlmUsedWhenCalculatingMa
uint32_t workGroupSize = static_cast<uint32_t>(lws[0] * lws[1] * lws[2]);
uint32_t expected = helper.calculateAvailableThreadCount(*rootDeviceEnvironment->getHardwareInfo(), grf) / static_cast<uint32_t>(Math::divideAndRoundUp(workGroupSize, simd));
expected = helper.adjustMaxWorkGroupCount(expected, EngineGroupType::compute, *rootDeviceEnvironment, false);
expected = helper.adjustMaxWorkGroupCount(expected, EngineGroupType::compute, *rootDeviceEnvironment);
EXPECT_EQ(expected, getMaxWorkGroupCount());
}
@@ -246,4 +246,4 @@ TEST_F(KernelHelperTest, GivenPtrByValueWhenCheckingIsAnyArgumentPtrByValueThenT
kernelDescriptor.payloadMappings.explicitArgs.push_back(pointerArg);
kernelDescriptor.payloadMappings.explicitArgs.push_back(valueArg);
EXPECT_TRUE(KernelHelper::isAnyArgumentPtrByValue(kernelDescriptor));
}
}

4
shared/test/unit_test/os_interface/linux/drm_debug_prelim_tests.cpp
View File

@@ -369,7 +369,7 @@ TEST_F(DrmDebugPrelimTest, givenAddedBindExtHandlesInBoWhenBindingWithinInternal
bo.addBindExtHandle(4);
bo.addBindExtHandle(5);
OsContextLinux osContext(drm, 0, 0u, {{aub_stream::EngineType::ENGINE_RCS, EngineUsage::internal}, 1 /*deviceBitfield*/, PreemptionMode::Disabled, true /* isRootDevice*/, false /* isEngineInstanced*/});
OsContextLinux osContext(drm, 0, 0u, {{aub_stream::EngineType::ENGINE_RCS, EngineUsage::internal}, 1 /*deviceBitfield*/, PreemptionMode::Disabled, true /* isRootDevice*/});
osContext.ensureContextInitialized(false);
bo.bind(&osContext, 0);
@@ -385,7 +385,7 @@ TEST_F(DrmDebugPrelimTest, givenAddedBindExtHandlesInBoWhenBindingWithinCopyEngi
drm.context.receivedVmBindUuidExt[0].reset();
OsContextLinux osContext(drm, 0, 0u, {{aub_stream::EngineType::ENGINE_BCS, EngineUsage::regular}, 1 /*deviceBitfield*/, PreemptionMode::Disabled, true /* isRootDevice*/, false /* isEngineInstanced*/});
OsContextLinux osContext(drm, 0, 0u, {{aub_stream::EngineType::ENGINE_BCS, EngineUsage::regular}, 1 /*deviceBitfield*/, PreemptionMode::Disabled, true /* isRootDevice*/});
osContext.ensureContextInitialized(false);
bo.bind(&osContext, 0);

36
shared/test/unit_test/os_interface/linux/drm_engine_info_prelim_tests.cpp
View File

@@ -176,7 +176,7 @@ static void givenEngineTypeWhenBindingDrmContextThenContextParamEngineIsSet(std:
ASSERT_NE(nullptr, drm->engineInfo);
auto drmContextId = 42u;
auto engineFlag = drm->bindDrmContext(drmContextId, 0u, engineType, false);
auto engineFlag = drm->bindDrmContext(drmContextId, 0u, engineType);
I915_DEFINE_CONTEXT_PARAM_ENGINES(enginesStruct, 1){};
EXPECT_EQ(static_cast<unsigned int>(I915_EXEC_DEFAULT), engineFlag);
@@ -206,7 +206,7 @@ TEST(DrmTest, givenRcsEngineWhenBindingDrmContextThenContextParamEngineIsSet) {
static void givenBcsEngineTypeWhenBindingDrmContextThenContextParamEngineIsSet(std::unique_ptr<DrmQueryMock> &drm, aub_stream::EngineType engineType, size_t numBcsSiblings, unsigned int engineIndex, uint32_t tileId) {
auto drmContextId = 42u;
drm->receivedContextParamRequestCount = 0u;
auto engineFlag = drm->bindDrmContext(drmContextId, tileId, engineType, false);
auto engineFlag = drm->bindDrmContext(drmContextId, tileId, engineType);
EXPECT_EQ(static_cast<unsigned int>(I915_EXEC_DEFAULT), engineFlag);
EXPECT_EQ(1u, drm->receivedContextParamRequestCount);
EXPECT_EQ(drmContextId, drm->receivedContextParamRequest.contextId);
@@ -305,7 +305,7 @@ HWCMDTEST_F(IGFX_XE_HP_CORE, DrmTestXeHPAndLater, givenLinkBcsEngineWithoutMainC
engineType = static_cast<aub_stream::EngineType>(aub_stream::ENGINE_BCS1 + engineIndex - 1);
numBcsSiblings -= 1;
}
auto engineFlag = drm->bindDrmContext(drmContextId, 0u, engineType, false);
auto engineFlag = drm->bindDrmContext(drmContextId, 0u, engineType);
if (engineIndex == 0) {
EXPECT_EQ(static_cast<unsigned int>(I915_EXEC_BLT), engineFlag);
@@ -358,7 +358,7 @@ HWTEST2_F(DrmTestXeHPAndLater, givenNotAllLinkBcsEnginesWhenBindingSingleTileDrm
auto drmContextId = 42u;
drm->receivedContextParamRequestCount = 0u;
auto engineFlag = drm->bindDrmContext(drmContextId, 0u, engineType, false);
auto engineFlag = drm->bindDrmContext(drmContextId, 0u, engineType);
if (drm->supportedCopyEnginesMask.test(engineIndex)) {
EXPECT_EQ(static_cast<unsigned int>(I915_EXEC_DEFAULT), engineFlag);
EXPECT_EQ(1u, drm->receivedContextParamRequestCount);
@@ -504,7 +504,7 @@ HWTEST2_F(DrmTestXeHPCAndLater, givenBcsVirtualEnginesDisabledWhenCreatingContex
drm->receivedContextParamRequestCount = 0u;
auto drmContextId = 42u;
auto engineType = static_cast<aub_stream::EngineType>(engineBase + engineIndex);
auto engineFlag = drm->bindDrmContext(drmContextId, 0u, engineType, false);
auto engineFlag = drm->bindDrmContext(drmContextId, 0u, engineType);
EXPECT_EQ(static_cast<unsigned int>(I915_EXEC_DEFAULT), engineFlag);
EXPECT_EQ(1u, drm->receivedContextParamRequestCount);
EXPECT_EQ(drmContextId, drm->receivedContextParamRequest.contextId);
@@ -546,7 +546,7 @@ TEST(DrmTest, givenVirtualEnginesEnabledWhenCreatingContextThenEnableLoadBalanci
drm->receivedContextParamRequestCount = 0u;
auto drmContextId = 42u;
auto engineType = static_cast<aub_stream::EngineType>(aub_stream::ENGINE_CCS + engineIndex);
auto engineFlag = drm->bindDrmContext(drmContextId, 0u, engineType, false);
auto engineFlag = drm->bindDrmContext(drmContextId, 0u, engineType);
EXPECT_EQ(static_cast<unsigned int>(I915_EXEC_DEFAULT), engineFlag);
EXPECT_EQ(1u, drm->receivedContextParamRequestCount);
EXPECT_EQ(drmContextId, drm->receivedContextParamRequest.contextId);
@@ -588,7 +588,7 @@ TEST(DrmTest, givenVirtualEnginesEnabledWhenCreatingContextThenEnableLoadBalanci
drm->receivedContextParamRequestCount = 0u;
auto drmContextId = 42u;
auto engineType = static_cast<aub_stream::EngineType>(aub_stream::ENGINE_CCS + engineIndex);
auto engineFlag = drm->bindDrmContext(drmContextId, 0u, engineType, false);
auto engineFlag = drm->bindDrmContext(drmContextId, 0u, engineType);
EXPECT_EQ(static_cast<unsigned int>(I915_EXEC_DEFAULT), engineFlag);
EXPECT_EQ(1u, drm->receivedContextParamRequestCount);
EXPECT_EQ(drmContextId, drm->receivedContextParamRequest.contextId);
@@ -654,7 +654,7 @@ TEST(DrmTest, givenVirtualEnginesDisabledWhenCreatingContextThenDontEnableLoadBa
drm->receivedContextParamRequestCount = 0u;
auto drmContextId = 42u;
auto engineType = static_cast<aub_stream::EngineType>(aub_stream::ENGINE_CCS + engineIndex);
auto engineFlag = drm->bindDrmContext(drmContextId, 0u, engineType, false);
auto engineFlag = drm->bindDrmContext(drmContextId, 0u, engineType);
EXPECT_EQ(static_cast<unsigned int>(I915_EXEC_DEFAULT), engineFlag);
EXPECT_EQ(1u, drm->receivedContextParamRequestCount);
EXPECT_EQ(drmContextId, drm->receivedContextParamRequest.contextId);
@@ -667,7 +667,7 @@ TEST(DrmTest, givenVirtualEnginesDisabledWhenCreatingContextThenDontEnableLoadBa
}
}
TEST(DrmTest, givenEngineInstancedDeviceWhenCreatingContextThenDontUseVirtualEngines) {
TEST(DrmTest, givenDeviceWhenCreatingContextThenDontUseVirtualEngines) {
auto executionEnvironment = std::make_unique<MockExecutionEnvironment>();
executionEnvironment->rootDeviceEnvironments[0]->initGmm();
@@ -685,16 +685,12 @@ TEST(DrmTest, givenEngineInstancedDeviceWhenCreatingContextThenDontUseVirtualEng
drm->receivedContextParamRequestCount = 0u;
auto drmContextId = 42u;
auto engineType = static_cast<aub_stream::EngineType>(aub_stream::ENGINE_CCS + engineIndex);
auto engineFlag = drm->bindDrmContext(drmContextId, 0u, engineType, true);
auto engineFlag = drm->bindDrmContext(drmContextId, 0u, engineType);
EXPECT_EQ(static_cast<unsigned int>(I915_EXEC_DEFAULT), engineFlag);
EXPECT_EQ(1u, drm->receivedContextParamRequestCount);
EXPECT_EQ(drmContextId, drm->receivedContextParamRequest.contextId);
EXPECT_EQ(static_cast<uint64_t>(I915_CONTEXT_PARAM_ENGINES), drm->receivedContextParamRequest.param);
EXPECT_EQ(ptrDiff(drm->receivedContextParamEngines.engines + 1, &drm->receivedContextParamEngines), drm->receivedContextParamRequest.size);
auto extensions = drm->receivedContextParamEngines.extensions;
EXPECT_EQ(0ull, extensions);
EXPECT_EQ(static_cast<__u16>(DrmPrelimHelper::getComputeEngineClass()), drm->receivedContextParamEngines.engines[0].engineClass);
EXPECT_EQ(static_cast<__u16>(engineIndex), DrmMockHelper::getIdFromEngineOrMemoryInstance(drm->receivedContextParamEngines.engines[0].engineInstance));
EXPECT_EQ(ptrDiff(drm->receivedContextParamEngines.engines + 5, &drm->receivedContextParamEngines), drm->receivedContextParamRequest.size);
}
}
@@ -712,7 +708,7 @@ TEST(DrmTest, givenVirtualEnginesEnabledAndNotEnoughCcsEnginesWhenCreatingContex
drm->receivedContextParamRequestCount = 0u;
auto drmContextId = 42u;
auto engineFlag = drm->bindDrmContext(drmContextId, 0u, aub_stream::ENGINE_CCS, false);
auto engineFlag = drm->bindDrmContext(drmContextId, 0u, aub_stream::ENGINE_CCS);
EXPECT_EQ(static_cast<unsigned int>(I915_EXEC_DEFAULT), engineFlag);
EXPECT_EQ(1u, drm->receivedContextParamRequestCount);
@@ -740,7 +736,7 @@ TEST(DrmTest, givenVirtualEnginesEnabledAndNonCcsEnginesWhenCreatingContextThenD
drm->receivedContextParamRequestCount = 0u;
auto drmContextId = 42u;
auto engineFlag = drm->bindDrmContext(drmContextId, 0u, defaultCopyEngine, false);
auto engineFlag = drm->bindDrmContext(drmContextId, 0u, defaultCopyEngine);
EXPECT_EQ(static_cast<unsigned int>(I915_EXEC_DEFAULT), engineFlag);
EXPECT_EQ(1u, drm->receivedContextParamRequestCount);
@@ -766,7 +762,7 @@ TEST(DrmTest, givenInvalidTileWhenBindingDrmContextThenErrorIsReturned) {
EXPECT_EQ(haveLocalMemory ? 3u : 2u, drm->ioctlCallsCount);
ASSERT_NE(nullptr, drm->engineInfo);
auto engineFlag = drm->bindDrmContext(42u, 20u, aub_stream::ENGINE_CCS, false);
auto engineFlag = drm->bindDrmContext(42u, 20u, aub_stream::ENGINE_CCS);
EXPECT_EQ(static_cast<unsigned int>(I915_EXEC_DEFAULT), engineFlag);
EXPECT_EQ(haveLocalMemory ? 3u : 2u, drm->ioctlCallsCount);
EXPECT_EQ(0u, drm->receivedContextParamRequestCount);
@@ -785,7 +781,7 @@ TEST(DrmTest, givenInvalidEngineTypeWhenBindingDrmContextThenExceptionIsThrown)
EXPECT_EQ(haveLocalMemory ? 3u : 2u, drm->ioctlCallsCount);
ASSERT_NE(nullptr, drm->engineInfo);
EXPECT_THROW(drm->bindDrmContext(42u, 0u, aub_stream::ENGINE_VCS, false), std::exception);
EXPECT_THROW(drm->bindDrmContext(42u, 0u, aub_stream::ENGINE_VCS), std::exception);
EXPECT_EQ(haveLocalMemory ? 3u : 2u, drm->ioctlCallsCount);
EXPECT_EQ(0u, drm->receivedContextParamRequestCount);
}
@@ -807,7 +803,7 @@ TEST(DrmTest, givenSetParamEnginesFailsWhenBindingDrmContextThenCallUnrecoverabl
drm->storedRetValForSetParamEngines = -1;
auto drmContextId = 42u;
EXPECT_ANY_THROW(drm->bindDrmContext(drmContextId, 0u, renderEngine, false));
EXPECT_ANY_THROW(drm->bindDrmContext(drmContextId, 0u, renderEngine));
}
TEST(DrmTest, whenQueryingEngineInfoThenMultiTileArchInfoIsUnchanged) {

6
shared/test/unit_test/os_interface/linux/xe/ioctl_helper_xe_tests.cpp
View File

@@ -1886,9 +1886,9 @@ TEST(IoctlHelperXeTest, whenBindingDrmContextWithoutVirtualEnginesThenProperEngi
uint16_t tileId = 1u;
uint16_t expectedGtId = 2u;
EXPECT_EQ(expectedValue, drm->bindDrmContext(0, tileId, aub_stream::EngineType::ENGINE_CCS, true));
EXPECT_EQ(expectedValue, drm->bindDrmContext(0, tileId, aub_stream::EngineType::ENGINE_CCS));
EXPECT_EQ(1u, ioctlHelper->contextParamEngine.size());
EXPECT_EQ(4u, ioctlHelper->contextParamEngine.size());
auto expectedEngine = drm->getEngineInfo()->getEngineInstance(1, aub_stream::EngineType::ENGINE_CCS);
auto notExpectedEngine = drm->getEngineInfo()->getEngineInstance(0, aub_stream::EngineType::ENGINE_CCS);
EXPECT_NE(expectedEngine->engineInstance, notExpectedEngine->engineInstance);
@@ -1910,7 +1910,7 @@ TEST(IoctlHelperXeTest, whenBindingDrmContextWithVirtualEnginesThenProperEngines
unsigned int expectedValue = DRM_XE_ENGINE_CLASS_COMPUTE;
uint16_t tileId = 1u;
uint16_t expectedGtId = 2u;
EXPECT_EQ(expectedValue, drm->bindDrmContext(0, tileId, aub_stream::EngineType::ENGINE_CCS, false));
EXPECT_EQ(expectedValue, drm->bindDrmContext(0, tileId, aub_stream::EngineType::ENGINE_CCS));
EXPECT_EQ(2u, ioctlHelper->contextParamEngine.size());
{

2
shared/test/unit_test/os_interface/windows/wddm_tests.cpp
View File

@@ -157,7 +157,7 @@ TEST(WddmNewRsourceTest, whenSetNewResourcesBoundToPageTableThenSetInContextFrom
executionEnvironment.initializeMemoryManager();
auto csr1 = std::unique_ptr<CommandStreamReceiver>(createCommandStream(executionEnvironment, 0, 1));
auto csr2 = std::unique_ptr<CommandStreamReceiver>(createCommandStream(executionEnvironment, 1, 1));
EngineDescriptor engineDesc({aub_stream::ENGINE_CCS, EngineUsage::regular}, 1, PreemptionMode::Disabled, false, false);
EngineDescriptor engineDesc({aub_stream::ENGINE_CCS, EngineUsage::regular}, 1, PreemptionMode::Disabled, false);
executionEnvironment.memoryManager->createAndRegisterOsContext(csr1.get(), engineDesc);
executionEnvironment.memoryManager->createAndRegisterOsContext(csr2.get(), engineDesc);

4
shared/test/unit_test/xe_hpc_core/pvc/gfx_core_helper_tests_pvc.cpp
View File

@@ -183,14 +183,14 @@ PVCTEST_F(GfxCoreHelperTestsPvc, GivenCooperativeEngineSupportedAndNotUsedWhenAd
(engineGroupType == EngineGroupType::compute && isRcsEnabled)) &&
productHelper.isCooperativeEngineSupported(hwInfo);
if (disallowDispatch) {
EXPECT_EQ(1u, gfxCoreHelper.adjustMaxWorkGroupCount(passedMaxWorkGroupCount, engineGroupType, rootDeviceEnvironment, false));
EXPECT_EQ(1u, gfxCoreHelper.adjustMaxWorkGroupCount(passedMaxWorkGroupCount, engineGroupType, rootDeviceEnvironment));
} else {
for (uint32_t ccsCount : {1, 2, 4}) {
hwInfo.gtSystemInfo.CCSInfo.NumberOfCCSEnabled = ccsCount;
tilePartsForConcurrentKernels = ccsCount == 1 ? 1
: ccsCount == 2 ? 4
: 8;
EXPECT_EQ(passedMaxWorkGroupCount / tilePartsForConcurrentKernels, gfxCoreHelper.adjustMaxWorkGroupCount(passedMaxWorkGroupCount, engineGroupType, rootDeviceEnvironment, false));
EXPECT_EQ(passedMaxWorkGroupCount / tilePartsForConcurrentKernels, gfxCoreHelper.adjustMaxWorkGroupCount(passedMaxWorkGroupCount, engineGroupType, rootDeviceEnvironment));
}
}
}

2
shared/test/unit_test/xe_hpc_core/pvc/test_preamble_pvc.cpp
View File

@@ -32,7 +32,7 @@ PVCTEST_F(PreambleCfeState, givenXeHpcAndKernelExecutionTypeAndRevisionWhenCalli
StreamProperties streamProperties{};
streamProperties.initSupport(pDevice->getRootDeviceEnvironment());
hwInfo->platform.usRevId = productHelper.getHwRevIdFromStepping(revision, *hwInfo);
streamProperties.frontEndState.setPropertiesAll(kernelExecutionType, false, false, false);
streamProperties.frontEndState.setPropertiesAll(kernelExecutionType, false, false);
PreambleHelper<FamilyType>::programVfeState(pVfeCmd, pDevice->getRootDeviceEnvironment(), 0u, 0, 0, streamProperties);
parseCommands<FamilyType>(linearStream);

2
shared/test/unit_test/xe_hpc_core/test_preamble_xe_hpc_core.cpp
View File

@@ -48,7 +48,7 @@ XE_HPC_CORETEST_F(PreambleCfeState, givenKernelExecutionTypeConcurrentAndRevisio
auto pVfeCmd = PreambleHelper<FamilyType>::getSpaceForVfeState(&linearStream, hwInfo, EngineGroupType::renderCompute);
StreamProperties streamProperties{};
streamProperties.initSupport(pDevice->getRootDeviceEnvironment());
streamProperties.frontEndState.setPropertiesAll(true, false, false, false);
streamProperties.frontEndState.setPropertiesAll(true, false, false);
PreambleHelper<FamilyType>::programVfeState(pVfeCmd, pDevice->getRootDeviceEnvironment(), 0u, 0, 0, streamProperties);
parseCommands<FamilyType>(linearStream);