refactor: remove redundant tracking of media sampler programming

Signed-off-by: Mateusz Jablonski <mateusz.jablonski@intel.com>

level_zero/core/source/cmdlist/cmdlist_imp.cpp

@@ -307,7 +307,7 @@ void CommandListImp::setStreamPropertiesDefaultSettings(NEO::StreamProperties &s
     }
 
     streamProperties.frontEndState.setPropertiesDisableOverdispatch(cmdListDefaultDisableOverdispatch, true);
-    streamProperties.pipelineSelect.setPropertiesModeSelectedMediaSamplerClockGate(cmdListDefaultPipelineSelectModeSelected, cmdListDefaultMediaSamplerClockGate, true);
+    streamProperties.pipelineSelect.setPropertiesModeSelected(cmdListDefaultPipelineSelectModeSelected, true);
 }
 
 void CommandListImp::enableInOrderExecution() {

level_zero/core/source/cmdlist/cmdlist_imp.h

@@ -62,7 +62,6 @@ struct CommandListImp : public CommandList {
     static constexpr bool cmdListDefaultCoherency = false;
     static constexpr bool cmdListDefaultDisableOverdispatch = true;
     static constexpr bool cmdListDefaultPipelineSelectModeSelected = true;
-    static constexpr bool cmdListDefaultMediaSamplerClockGate = false;
     static constexpr bool cmdListDefaultGlobalAtomics = false;
     std::vector<Event *> mappedTsEventList;
     std::vector<Event *> interruptEvents;

level_zero/core/source/cmdqueue/cmdqueue_hw.inl

@@ -656,7 +656,7 @@ template <GFXCORE_FAMILY gfxCoreFamily>
 void CommandQueueHw<gfxCoreFamily>::programPipelineSelectIfGpgpuDisabled(NEO::LinearStream &cmdStream) {
     bool gpgpuEnabled = this->csr->getPreambleSetFlag();
     if (!gpgpuEnabled) {
-        NEO::PipelineSelectArgs args = {false, false, false, false};
+        NEO::PipelineSelectArgs args = {false, false, false};
         NEO::PreambleHelper<GfxFamily>::programPipelineSelect(&cmdStream, args, device->getNEODevice()->getRootDeviceEnvironment());
         this->csr->setPreambleSetFlag(true);
     }
@@ -1566,7 +1566,6 @@ void CommandQueueHw<gfxCoreFamily>::programOneCmdListPipelineSelect(NEO::LinearS
         NEO::PipelineSelectArgs args = {
             systolic,
             false,
-            false,
             cmdListRequired.commandList->getSystolicModeSupport()};
 
         NEO::PreambleHelper<GfxFamily>::programPipelineSelect(&commandStream, args, device->getNEODevice()->getRootDeviceEnvironment());
@@ -1624,7 +1623,6 @@ void CommandQueueHw<gfxCoreFamily>::programRequiredStateComputeModeForCommandLis
         NEO::PipelineSelectArgs pipelineSelectArgs = {
             cmdListRequired.requiredState.pipelineSelect.systolicMode.value == 1,
             false,
-            false,
             cmdListRequired.commandList->getSystolicModeSupport()};
 
         NEO::EncodeComputeMode<GfxFamily>::programComputeModeCommandWithSynchronization(commandStream, cmdListRequired.requiredState.stateComputeMode, pipelineSelectArgs,

level_zero/core/test/unit_tests/sources/cmdlist/test_cmdlist_6.cpp

@@ -274,7 +274,6 @@ HWTEST_F(CommandListExecuteImmediate, GivenImmediateCommandListWhenCommandListIs
     EXPECT_EQ(-1, currentCsrStreamProperties.frontEndState.singleSliceDispatchCcsMode.value);
 
     EXPECT_EQ(-1, currentCsrStreamProperties.pipelineSelect.modeSelected.value);
-    EXPECT_EQ(-1, currentCsrStreamProperties.pipelineSelect.mediaSamplerDopClockGate.value);
 }
 
 struct CommandListTest : Test<DeviceFixture> {

opencl/test/unit_test/command_queue/enqueue_kernel_1_tests.cpp

@@ -1728,17 +1728,6 @@ HWTEST_TEMPLATED_F(EnqueueKernelTestWithMockCsrHw2, givenContextWithSeveralDevic
     context->deviceBitfields[rootDeviceIndex].set(3, false);
 }
 
-HWTEST_TEMPLATED_F(EnqueueKernelTestWithMockCsrHw2, givenNonVMEKernelWhenEnqueueKernelThenDispatchFlagsDoesntHaveMediaSamplerRequired) {
-    auto *mockCsr = static_cast<MockCsrHw2<FamilyType> *>(&pDevice->getGpgpuCommandStreamReceiver());
-    mockCsr->overrideDispatchPolicy(DispatchMode::batchedDispatch);
-
-    MockKernelWithInternals mockKernel(*pClDevice, context);
-    size_t gws[3] = {1, 0, 0};
-    mockKernel.kernelInfo.kernelDescriptor.kernelAttributes.flags.usesVme = false;
-    clEnqueueNDRangeKernel(this->pCmdQ, mockKernel.mockMultiDeviceKernel, 1, nullptr, gws, nullptr, 0, nullptr, nullptr);
-    EXPECT_FALSE(mockCsr->passedDispatchFlags.pipelineSelectArgs.mediaSamplerRequired);
-}
-
 HWTEST_F(EnqueueKernelTest, whenEnqueueKernelWithEngineHintsThenEpilogRequiredIsSet) {
     auto &csr = pDevice->getUltCommandStreamReceiver<FamilyType>();
     size_t off[3] = {0, 0, 0};

opencl/test/unit_test/command_stream/command_stream_receiver_flush_task_1_tests.cpp

@@ -718,28 +718,17 @@ HWTEST_F(CommandStreamReceiverFlushTaskTests, givenFlushTaskWhenInitProgrammingF
     EXPECT_FALSE(commandStreamReceiver.bindingTableBaseAddressRequired);
 }
 
-HWTEST2_F(CommandStreamReceiverFlushTaskTests, GivenPreambleNotSentAndMediaSamplerRequirementChangedWhenFlushingTaskThenPipelineSelectIsSent, IsAtMostXeCore) {
+HWTEST2_F(CommandStreamReceiverFlushTaskTests, GivenPreambleNotSentWhenFlushingTaskThenPipelineSelectIsSent, IsAtMostXeCore) {
     auto &commandStreamReceiver = pDevice->getUltCommandStreamReceiver<FamilyType>();
     commandStreamReceiver.isPreambleSent = false;
-    commandStreamReceiver.lastMediaSamplerConfig = -1;
     flushTask(commandStreamReceiver);
     parseCommands<FamilyType>(commandStreamReceiver.commandStream, 0);
     EXPECT_NE(nullptr, getCommand<typename FamilyType::PIPELINE_SELECT>());
 }
 
-HWTEST2_F(CommandStreamReceiverFlushTaskTests, GivenPreambleNotSentAndMediaSamplerRequirementNotChangedWhenFlushingTaskThenPipelineSelectIsSent, IsAtMostXeCore) {
+HWTEST_F(CommandStreamReceiverFlushTaskTests, GivenPreambleSentWhenFlushingTaskThenPipelineSelectIsNotSent) {
-    auto &commandStreamReceiver = pDevice->getUltCommandStreamReceiver<FamilyType>();
-    commandStreamReceiver.isPreambleSent = false;
-    commandStreamReceiver.lastMediaSamplerConfig = 0;
-    flushTask(commandStreamReceiver);
-    parseCommands<FamilyType>(commandStreamReceiver.commandStream, 0);
-    EXPECT_NE(nullptr, getCommand<typename FamilyType::PIPELINE_SELECT>());
-}
-
-HWTEST_F(CommandStreamReceiverFlushTaskTests, GivenPreambleSentAndMediaSamplerRequirementNotChangedWhenFlushingTaskThenPipelineSelectIsNotSent) {
     auto &commandStreamReceiver = pDevice->getUltCommandStreamReceiver<FamilyType>();
     commandStreamReceiver.isPreambleSent = true;
-    commandStreamReceiver.lastMediaSamplerConfig = 0;
     flushTask(commandStreamReceiver);
     parseCommands<FamilyType>(commandStreamReceiver.commandStream, 0);
     auto &productHelper = pDevice->getProductHelper();
@@ -749,15 +738,6 @@ HWTEST_F(CommandStreamReceiverFlushTaskTests, GivenPreambleSentAndMediaSamplerRe
         EXPECT_EQ(nullptr, getCommand<typename FamilyType::PIPELINE_SELECT>());
     }
 }
-HWTEST2_F(CommandStreamReceiverFlushTaskTests, GivenPreambleSentAndMediaSamplerRequirementChangedWhenFlushingTaskThenPipelineSelectIsSent, IsAtMostXeCore) {
-    auto &commandStreamReceiver = pDevice->getUltCommandStreamReceiver<FamilyType>();
-    commandStreamReceiver.pipelineSupportFlags.mediaSamplerDopClockGate = true;
-    commandStreamReceiver.isPreambleSent = true;
-    commandStreamReceiver.lastMediaSamplerConfig = 1;
-    flushTask(commandStreamReceiver);
-    parseCommands<FamilyType>(commandStreamReceiver.commandStream, 0);
-    EXPECT_NE(nullptr, getCommand<typename FamilyType::PIPELINE_SELECT>());
-}
 
 HWTEST2_F(CommandStreamReceiverFlushTaskTests, GivenStateBaseAddressNotSentWhenFlushingTaskThenStateBaseAddressIsSent, IsHeapfulSupported) {
     auto &commandStreamReceiver = pDevice->getUltCommandStreamReceiver<FamilyType>();

opencl/test/unit_test/command_stream/command_stream_receiver_flush_task_2_tests.cpp

@@ -1064,7 +1064,6 @@ HWCMDTEST_F(IGFX_GEN12LP_CORE, CommandStreamReceiverFlushTaskTests, GivenPreambl
     taskLevel = commandStreamReceiver.peekTaskLevel() + 1;
     commandStreamReceiver.isPreambleSent = true;
     commandStreamReceiver.lastPreemptionMode = pDevice->getPreemptionMode();
-    commandStreamReceiver.lastMediaSamplerConfig = 0;
     commandStreamReceiver.streamProperties.stateComputeMode.isCoherencyRequired.value = 0;
     csrSizeRequest.l3ConfigChanged = true;
     commandStreamReceiver.overrideCsrSizeReqFlags(csrSizeRequest);
@@ -1117,24 +1116,6 @@ HWTEST_F(CommandStreamReceiverFlushTaskTests, givenCsrWhenPreambleNotSentThenReq
     EXPECT_EQ(l3ConfigNotChangedSize, l3ConfigChangedSize);
 }
 
-HWTEST_F(CommandStreamReceiverFlushTaskTests, givenCsrWhenPreambleNotSentThenRequiredCsrSizeDoesntDependOnmediaSamplerConfigChanged) {
-    UltCommandStreamReceiver<FamilyType> &commandStreamReceiver = (UltCommandStreamReceiver<FamilyType> &)pDevice->getGpgpuCommandStreamReceiver();
-    CsrSizeRequestFlags csrSizeRequest = {};
-    DispatchFlags flags = DispatchFlagsHelper::createDefaultDispatchFlags();
-
-    commandStreamReceiver.isPreambleSent = false;
-
-    csrSizeRequest.mediaSamplerConfigChanged = false;
-    commandStreamReceiver.overrideCsrSizeReqFlags(csrSizeRequest);
-    auto mediaSamplerConfigNotChangedSize = commandStreamReceiver.getRequiredCmdStreamSize(flags, *pDevice);
-
-    csrSizeRequest.mediaSamplerConfigChanged = true;
-    commandStreamReceiver.overrideCsrSizeReqFlags(csrSizeRequest);
-    auto mediaSamplerConfigChangedSize = commandStreamReceiver.getRequiredCmdStreamSize(flags, *pDevice);
-
-    EXPECT_EQ(mediaSamplerConfigChangedSize, mediaSamplerConfigNotChangedSize);
-}
-
 HWTEST_F(CommandStreamReceiverFlushTaskTests, givenCsrWhenSamplerCacheFlushSentThenRequiredCsrSizeContainsPipecontrolSize) {
     UltCommandStreamReceiver<FamilyType> &commandStreamReceiver = (UltCommandStreamReceiver<FamilyType> &)pDevice->getGpgpuCommandStreamReceiver();
     CsrSizeRequestFlags csrSizeRequest = {};

opencl/test/unit_test/fixtures/ult_command_stream_receiver_fixture.h

@@ -165,8 +165,7 @@ struct UltCommandStreamReceiverTest
         configureCSRHeapStatesToNonDirty<GfxFamily>();
         commandStreamReceiver.taskLevel = taskLevel;
 
-        commandStreamReceiver.lastMediaSamplerConfig = 0;
+        commandStreamReceiver.streamProperties.pipelineSelect.setPropertiesAll(true, false);
-        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);

opencl/test/unit_test/xe2_hpg_core/command_stream_receiver_hw_tests_xe2_hpg_core.cpp

@@ -554,8 +554,6 @@ XE2_HPG_CORETEST_F(Xe2CommandStreamReceiverFlushTaskTests, givenNotExistPolicyWh
 XE2_HPG_CORETEST_F(Xe2CommandStreamReceiverFlushTaskTests, givenLastSystolicPipelineSelectModeWhenFlushTaskIsCalledThenDontReprogramPipelineSelect) {
     auto &commandStreamReceiver = pDevice->getUltCommandStreamReceiver<FamilyType>();
     commandStreamReceiver.isPreambleSent = true;
-    commandStreamReceiver.lastMediaSamplerConfig = false;
-    flushTaskFlags.pipelineSelectArgs.mediaSamplerRequired = false;
     flushTaskFlags.pipelineSelectArgs.systolicPipelineSelectMode = true;
 
     flushTask(commandStreamReceiver);

opencl/test/unit_test/xe3_core/command_stream_receiver_hw_tests_xe3_core.cpp

@@ -325,8 +325,6 @@ XE3_CORETEST_F(Xe3CommandStreamReceiverFlushTaskTests, givenNotExistPolicyWhenFl
 XE3_CORETEST_F(Xe3CommandStreamReceiverFlushTaskTests, givenLastSystolicPipelineSelectModeWhenFlushTaskIsCalledThenDontReprogramPipelineSelect) {
     auto &commandStreamReceiver = pDevice->getUltCommandStreamReceiver<FamilyType>();
     commandStreamReceiver.isPreambleSent = true;
-    commandStreamReceiver.lastMediaSamplerConfig = false;
-    flushTaskFlags.pipelineSelectArgs.mediaSamplerRequired = false;
     flushTaskFlags.pipelineSelectArgs.systolicPipelineSelectMode = true;
 
     flushTask(commandStreamReceiver);

opencl/test/unit_test/xe_hpc_core/pvc/command_stream_receiver_hw_tests_pvc.cpp

@@ -71,7 +71,6 @@ PVCTEST_F(PvcCommandStreamReceiverFlushTaskTests, givenRevisionBAndAboveWhenLast
     auto &commandStreamReceiver = pDevice->getUltCommandStreamReceiver<FamilyType>();
 
     flushTaskFlags.pipelineSelectArgs.systolicPipelineSelectMode = true;
-    flushTaskFlags.pipelineSelectArgs.mediaSamplerRequired = false;
 
     struct {
         unsigned short revId;
@@ -89,7 +88,6 @@ PVCTEST_F(PvcCommandStreamReceiverFlushTaskTests, givenRevisionBAndAboveWhenLast
         hwInfo->platform.usRevId = testInput.revId;
         productHelper.fillPipelineSelectPropertiesSupportStructure(commandStreamReceiver.pipelineSupportFlags, *hwInfo);
         commandStreamReceiver.isPreambleSent = true;
-        commandStreamReceiver.lastMediaSamplerConfig = false;
         commandStreamReceiver.lastSystolicPipelineSelectMode = false;
         commandStreamReceiver.streamProperties.pipelineSelect.systolicMode.value = -1;
 

shared/source/command_stream/command_stream_receiver.cpp

@@ -599,7 +599,6 @@ void CommandStreamReceiver::initProgrammingFlags() {
     stateComputeModeDirty = true;
 
     lastSentL3Config = 0;
-    lastMediaSamplerConfig = -1;
     lastPreemptionMode = PreemptionMode::Initial;
 
     latestSentStatelessMocsConfig = CacheSettings::unknownMocs;

shared/source/command_stream/command_stream_receiver.h

@@ -674,8 +674,6 @@ class CommandStreamReceiver : NEO::NonCopyableAndNonMovableClass {
     const uint32_t rootDeviceIndex;
     const DeviceBitfield deviceBitfield;
 
-    int8_t lastMediaSamplerConfig = -1;
-
     bool isPreambleSent = false;
     bool isStateSipSent = false;
     bool isEnginePrologueSent = false;

shared/source/command_stream/command_stream_receiver_hw_base.inl

@@ -849,8 +849,7 @@ size_t CommandStreamReceiverHw<GfxFamily>::getRequiredCmdStreamSize(const Dispat
 template <typename GfxFamily>
 inline size_t CommandStreamReceiverHw<GfxFamily>::getCmdSizeForPipelineSelect() const {
     size_t size = 0;
-    if ((csrSizeRequestFlags.mediaSamplerConfigChanged ||
+    if ((csrSizeRequestFlags.systolicPipelineSelectMode ||
-         csrSizeRequestFlags.systolicPipelineSelectMode ||
          !isPreambleSent) &&
         !isPipelineSelectAlreadyProgrammed()) {
         size += PreambleHelper<GfxFamily>::getCmdSizeForPipelineSelect(peekRootDeviceEnvironment());
@@ -1571,15 +1570,10 @@ void CommandStreamReceiverHw<GfxFamily>::handleFrontEndStateTransition(const Dis
 
 template <typename GfxFamily>
 void CommandStreamReceiverHw<GfxFamily>::handlePipelineSelectStateTransition(const DispatchFlags &dispatchFlags) {
-    if (streamProperties.pipelineSelect.mediaSamplerDopClockGate.value != -1) {
-        this->lastMediaSamplerConfig = static_cast<int8_t>(streamProperties.pipelineSelect.mediaSamplerDopClockGate.value);
-    }
     if (streamProperties.pipelineSelect.systolicMode.value != -1) {
         this->lastSystolicPipelineSelectMode = !!streamProperties.pipelineSelect.systolicMode.value;
     }
 
-    csrSizeRequestFlags.mediaSamplerConfigChanged = this->pipelineSupportFlags.mediaSamplerDopClockGate &&
-                                                    (this->lastMediaSamplerConfig != static_cast<int8_t>(dispatchFlags.pipelineSelectArgs.mediaSamplerRequired));
     csrSizeRequestFlags.systolicPipelineSelectMode = this->pipelineSupportFlags.systolicMode &&
                                                      (this->lastSystolicPipelineSelectMode != dispatchFlags.pipelineSelectArgs.systolicPipelineSelectMode);
 }
@@ -2005,7 +1999,6 @@ void CommandStreamReceiverHw<GfxFamily>::handleImmediateFlushPipelineSelectState
     flushData.pipelineSelectArgs = {
         this->streamProperties.pipelineSelect.systolicMode.value == 1,
         false,
-        false,
         this->pipelineSupportFlags.systolicMode};
 }
 

shared/source/command_stream/command_stream_receiver_hw_xehp_and_later.inl

@@ -44,11 +44,10 @@ size_t CommandStreamReceiverHw<GfxFamily>::getCmdSizeForL3Config() const { retur
 
 template <typename GfxFamily>
 void CommandStreamReceiverHw<GfxFamily>::programPipelineSelect(LinearStream &commandStream, PipelineSelectArgs &pipelineSelectArgs) {
-    if (csrSizeRequestFlags.mediaSamplerConfigChanged || csrSizeRequestFlags.systolicPipelineSelectMode || !isPreambleSent) {
+    if (csrSizeRequestFlags.systolicPipelineSelectMode || !isPreambleSent) {
         PreambleHelper<GfxFamily>::programPipelineSelect(&commandStream, pipelineSelectArgs, peekRootDeviceEnvironment());
-        this->lastMediaSamplerConfig = pipelineSelectArgs.mediaSamplerRequired;
         this->lastSystolicPipelineSelectMode = pipelineSelectArgs.systolicPipelineSelectMode;
-        this->streamProperties.pipelineSelect.setPropertiesAll(true, this->lastMediaSamplerConfig, this->lastSystolicPipelineSelectMode);
+        this->streamProperties.pipelineSelect.setPropertiesAll(true, this->lastSystolicPipelineSelectMode);
         this->streamProperties.pipelineSelect.clearIsDirty();
     }
 }

shared/source/command_stream/csr_definitions.h

@@ -133,7 +133,6 @@ struct DispatchFlags {
 struct CsrSizeRequestFlags {
     bool l3ConfigChanged = false;
     bool preemptionRequestChanged = false;
-    bool mediaSamplerConfigChanged = false;
     bool hasSharedHandles = false;
     bool systolicPipelineSelectMode = false;
     bool activePartitionsChanged = false;

shared/source/command_stream/definitions/stream_properties.inl

@@ -102,34 +102,6 @@ struct FrontEndProperties {
     bool propertiesSupportLoaded = false;
 };
 
-struct PipelineSelectPropertiesSupport {
-    bool mediaSamplerDopClockGate = false;
-    bool systolicMode = false;
-};
-
-struct PipelineSelectProperties {
-    StreamProperty modeSelected{};
-    StreamProperty mediaSamplerDopClockGate{};
-    StreamProperty systolicMode{};
-
-    void initSupport(const RootDeviceEnvironment &rootDeviceEnvironment);
-    void resetState();
-
-    void setPropertiesAll(bool modeSelected, bool mediaSamplerDopClockGate, bool systolicMode);
-    void setPropertiesModeSelectedMediaSamplerClockGate(bool modeSelected, bool mediaSamplerDopClockGate, bool clearDirtyState);
-    void setPropertySystolicMode(bool systolicMode);
-
-    void copyPropertiesAll(const PipelineSelectProperties &properties);
-    void copyPropertiesSystolicMode(const PipelineSelectProperties &properties);
-
-    bool isDirty() const;
-    void clearIsDirty();
-
-  protected:
-    PipelineSelectPropertiesSupport pipelineSelectPropertiesSupport = {};
-    bool propertiesSupportLoaded = false;
-};
-
 struct StateBaseAddressPropertiesSupport {
     bool bindingTablePoolBaseAddress = false;
 };

shared/source/command_stream/stream_properties.cpp

@@ -310,42 +310,31 @@ void PipelineSelectProperties::resetState() {
     clearIsDirty();
 
     this->modeSelected.value = StreamProperty::initValue;
-    this->mediaSamplerDopClockGate.value = StreamProperty::initValue;
     this->systolicMode.value = StreamProperty::initValue;
 }
 
-void PipelineSelectProperties::setPropertiesAll(bool modeSelected, bool mediaSamplerDopClockGate, bool systolicMode) {
+void PipelineSelectProperties::setPropertiesAll(bool modeSelected, bool systolicMode) {
     DEBUG_BREAK_IF(!this->propertiesSupportLoaded);
     clearIsDirty();
 
     this->modeSelected.set(modeSelected);
 
-    if (this->pipelineSelectPropertiesSupport.mediaSamplerDopClockGate) {
-        this->mediaSamplerDopClockGate.set(mediaSamplerDopClockGate);
-    }
-
     if (this->pipelineSelectPropertiesSupport.systolicMode) {
         this->systolicMode.set(systolicMode);
     }
 }
 
-void PipelineSelectProperties::setPropertiesModeSelectedMediaSamplerClockGate(bool modeSelected, bool mediaSamplerDopClockGate, bool clearDirtyState) {
+void PipelineSelectProperties::setPropertiesModeSelected(bool modeSelected, bool clearDirtyState) {
     DEBUG_BREAK_IF(!this->propertiesSupportLoaded);
 
     if (!clearDirtyState) {
         this->modeSelected.isDirty = false;
-        this->mediaSamplerDopClockGate.isDirty = false;
     }
 
     this->modeSelected.set(modeSelected);
 
-    if (this->pipelineSelectPropertiesSupport.mediaSamplerDopClockGate) {
-        this->mediaSamplerDopClockGate.set(mediaSamplerDopClockGate);
-    }
-
     if (clearDirtyState) {
         this->modeSelected.isDirty = false;
-        this->mediaSamplerDopClockGate.isDirty = false;
     }
 }
 
@@ -363,7 +352,6 @@ void PipelineSelectProperties::copyPropertiesAll(const PipelineSelectProperties
     clearIsDirty();
 
     modeSelected.set(properties.modeSelected.value);
-    mediaSamplerDopClockGate.set(properties.mediaSamplerDopClockGate.value);
     systolicMode.set(properties.systolicMode.value);
 }
 
@@ -373,12 +361,11 @@ void PipelineSelectProperties::copyPropertiesSystolicMode(const PipelineSelectPr
 }
 
 bool PipelineSelectProperties::isDirty() const {
-    return modeSelected.isDirty || mediaSamplerDopClockGate.isDirty || systolicMode.isDirty;
+    return modeSelected.isDirty || systolicMode.isDirty;
 }
 
 void PipelineSelectProperties::clearIsDirty() {
     modeSelected.isDirty = false;
-    mediaSamplerDopClockGate.isDirty = false;
     systolicMode.isDirty = false;
 }
 

shared/source/command_stream/stream_properties.h

@@ -1,5 +1,5 @@
 /*
- * Copyright (C) 2021-2023 Intel Corporation
+ * Copyright (C) 2021-2025 Intel Corporation
  *
  * SPDX-License-Identifier: MIT
  *
@@ -11,6 +11,32 @@
 
 namespace NEO {
 
+struct PipelineSelectPropertiesSupport {
+    bool systolicMode = false;
+};
+
+struct PipelineSelectProperties {
+    StreamProperty modeSelected{};
+    StreamProperty systolicMode{};
+
+    void initSupport(const RootDeviceEnvironment &rootDeviceEnvironment);
+    void resetState();
+
+    void setPropertiesAll(bool modeSelected, bool systolicMode);
+    void setPropertiesModeSelected(bool modeSelected, bool clearDirtyState);
+    void setPropertySystolicMode(bool systolicMode);
+
+    void copyPropertiesAll(const PipelineSelectProperties &properties);
+    void copyPropertiesSystolicMode(const PipelineSelectProperties &properties);
+
+    bool isDirty() const;
+    void clearIsDirty();
+
+  protected:
+    PipelineSelectPropertiesSupport pipelineSelectPropertiesSupport = {};
+    bool propertiesSupportLoaded = false;
+};
+
 struct StreamProperties {
     StateComputeModeProperties stateComputeMode{};
     FrontEndProperties frontEndState{};

shared/source/gen12lp/command_stream_receiver_hw_gen12lp.cpp

@@ -39,13 +39,12 @@ size_t CommandStreamReceiverHw<GfxFamily>::getRequiredStateBaseAddressSize(const
 
 template <typename GfxFamily>
 void CommandStreamReceiverHw<GfxFamily>::programPipelineSelect(LinearStream &commandStream, PipelineSelectArgs &pipelineSelectArgs) {
-    if (csrSizeRequestFlags.mediaSamplerConfigChanged || csrSizeRequestFlags.systolicPipelineSelectMode || !isPreambleSent) {
+    if (csrSizeRequestFlags.systolicPipelineSelectMode || !isPreambleSent) {
         if (!isPipelineSelectAlreadyProgrammed()) {
             PreambleHelper<GfxFamily>::programPipelineSelect(&commandStream, pipelineSelectArgs, peekRootDeviceEnvironment());
         }
-        this->lastMediaSamplerConfig = pipelineSelectArgs.mediaSamplerRequired;
         this->lastSystolicPipelineSelectMode = pipelineSelectArgs.systolicPipelineSelectMode;
-        this->streamProperties.pipelineSelect.setPropertiesAll(true, this->lastMediaSamplerConfig, this->lastSystolicPipelineSelectMode);
+        this->streamProperties.pipelineSelect.setPropertiesAll(true, this->lastSystolicPipelineSelectMode);
         this->streamProperties.pipelineSelect.clearIsDirty();
     }
 }

shared/source/gen12lp/hw_cmds_adlp.h

@@ -1,5 +1,5 @@
 /*
- * Copyright (C) 2021-2023 Intel Corporation
+ * Copyright (C) 2021-2025 Intel Corporation
  *
  * SPDX-License-Identifier: MIT
  *
@@ -26,7 +26,6 @@ struct ADLP : public Gen12LpFamily {
     static void setupHardwareInfoBase(HardwareInfo *hwInfo, bool setupFeatureTableAndWorkaroundTable, const ReleaseHelper *releaseHelper);
 
     struct PipelineSelectStateSupport {
-        static constexpr bool mediaSamplerDopClockGate = true;
         static constexpr bool systolicMode = true;
     };
 };

shared/source/gen12lp/hw_cmds_base.h

@@ -53,7 +53,6 @@ struct Gen12Lp {
     };
 
     struct PipelineSelectStateSupport {
-        static constexpr bool mediaSamplerDopClockGate = true;
         static constexpr bool systolicMode = false;
     };
 

shared/source/gen12lp/preamble_gen12lp.cpp

@@ -105,7 +105,7 @@ void PreambleHelper<Family>::programPipelineSelect(LinearStream *pCommandStream,
     auto pipeline = pipelineSelectArgs.is3DPipelineRequired ? PIPELINE_SELECT::PIPELINE_SELECTION_3D : PIPELINE_SELECT::PIPELINE_SELECTION_GPGPU;
 
     pipelineSelectCmd.setPipelineSelection(pipeline);
-    pipelineSelectCmd.setMediaSamplerDopClockGateEnable(!pipelineSelectArgs.mediaSamplerRequired);
+    pipelineSelectCmd.setMediaSamplerDopClockGateEnable(true);
 
     if (pipelineSelectArgs.systolicPipelineSelectSupport) {
         mask |= pipelineSelectSystolicModeEnableMaskBits;

shared/source/helpers/pipeline_select_args.h

@@ -1,5 +1,5 @@
 /*
- * Copyright (C) 2019-2022 Intel Corporation
+ * Copyright (C) 2019-2025 Intel Corporation
  *
  * SPDX-License-Identifier: MIT
  *
@@ -10,7 +10,6 @@
 namespace NEO {
 struct PipelineSelectArgs {
     bool systolicPipelineSelectMode = false;
-    bool mediaSamplerRequired = false;
     bool is3DPipelineRequired = false;
     bool systolicPipelineSelectSupport = false;
 };

shared/source/os_interface/product_helper.h

@@ -213,7 +213,6 @@ class ProductHelper {
     virtual bool getPreemptionDbgPropertyStateSipSupport() const = 0;
     virtual bool getPreemptionDbgPropertyCsrSurfaceSupport() const = 0;
 
-    virtual bool getPipelineSelectPropertyMediaSamplerDopClockGateSupport() const = 0;
     virtual bool getPipelineSelectPropertySystolicModeSupport() const = 0;
 
     virtual void fillScmPropertiesSupportStructure(StateComputeModePropertiesSupport &propertiesSupport) const = 0;

shared/source/os_interface/product_helper.inl

@@ -779,12 +779,6 @@ void ProductHelperHw<gfxProduct>::fillFrontEndPropertiesSupportStructure(FrontEn
     propertiesSupport.singleSliceDispatchCcsMode = getFrontEndPropertySingleSliceDispatchCcsModeSupport();
 }
 
-template <PRODUCT_FAMILY gfxProduct>
-bool ProductHelperHw<gfxProduct>::getPipelineSelectPropertyMediaSamplerDopClockGateSupport() const {
-    using GfxProduct = typename HwMapper<gfxProduct>::GfxProduct;
-    return GfxProduct::PipelineSelectStateSupport::mediaSamplerDopClockGate;
-}
-
 template <PRODUCT_FAMILY gfxProduct>
 bool ProductHelperHw<gfxProduct>::getPipelineSelectPropertySystolicModeSupport() const {
     using GfxProduct = typename HwMapper<gfxProduct>::GfxProduct;
@@ -793,7 +787,6 @@ bool ProductHelperHw<gfxProduct>::getPipelineSelectPropertySystolicModeSupport()
 
 template <PRODUCT_FAMILY gfxProduct>
 void ProductHelperHw<gfxProduct>::fillPipelineSelectPropertiesSupportStructure(PipelineSelectPropertiesSupport &propertiesSupport, const HardwareInfo &hwInfo) const {
-    propertiesSupport.mediaSamplerDopClockGate = getPipelineSelectPropertyMediaSamplerDopClockGateSupport();
     propertiesSupport.systolicMode = isSystolicModeConfigurable(hwInfo);
 }
 

shared/source/os_interface/product_helper_hw.h

@@ -154,7 +154,6 @@ class ProductHelperHw : public ProductHelper {
     bool getPreemptionDbgPropertyStateSipSupport() const override;
     bool getPreemptionDbgPropertyCsrSurfaceSupport() const override;
 
-    bool getPipelineSelectPropertyMediaSamplerDopClockGateSupport() const override;
     bool getPipelineSelectPropertySystolicModeSupport() const override;
 
     void fillScmPropertiesSupportStructure(StateComputeModePropertiesSupport &propertiesSupport) const override;

shared/source/xe2_hpg_core/hw_cmds_base.h

@@ -60,7 +60,6 @@ struct Xe2HpgCore {
     };
 
     struct PipelineSelectStateSupport {
-        static constexpr bool mediaSamplerDopClockGate = false;
         static constexpr bool systolicMode = false;
     };
 

shared/source/xe3_core/hw_cmds_base.h

@@ -62,7 +62,6 @@ struct Xe3Core {
     };
 
     struct PipelineSelectStateSupport {
-        static constexpr bool mediaSamplerDopClockGate = false;
         static constexpr bool systolicMode = false;
     };
 

shared/source/xe_hpc_core/hw_cmds_xe_hpc_core_base.h

@@ -42,7 +42,6 @@ struct XeHpcCore {
     };
 
     struct PipelineSelectStateSupport {
-        static constexpr bool mediaSamplerDopClockGate = false;
         static constexpr bool systolicMode = true;
     };
 

shared/source/xe_hpg_core/hw_cmds_xe_hpg_core_base.h

@@ -60,7 +60,6 @@ struct XeHpgCore {
     };
 
     struct PipelineSelectStateSupport {
-        static constexpr bool mediaSamplerDopClockGate = false;
         static constexpr bool systolicMode = true;
     };
 

shared/test/common/fixtures/command_stream_receiver_fixture.inl

@@ -1,5 +1,5 @@
 /*
- * Copyright (C) 2022-2023 Intel Corporation
+ * Copyright (C) 2022-2025 Intel Corporation
  *
  * SPDX-License-Identifier: MIT
  *
@@ -20,10 +20,8 @@ void CommandStreamReceiverSystolicFixture::testBody() {
     bool systolicModeSupported = commandStreamReceiver.pipelineSupportFlags.systolicMode;
 
     commandStreamReceiver.isPreambleSent = true;
-    commandStreamReceiver.lastMediaSamplerConfig = false;
 
     flushTaskFlags.pipelineSelectArgs.systolicPipelineSelectMode = true;
-    flushTaskFlags.pipelineSelectArgs.mediaSamplerRequired = false;
 
     commandStreamReceiver.flushTask(commandStream,
                                     0,

shared/test/common/libult/ult_command_stream_receiver.h

@@ -140,7 +140,6 @@ class UltCommandStreamReceiver : public CommandStreamReceiverHw<GfxFamily> {
     using BaseClass::CommandStreamReceiver::isStateSipSent;
     using BaseClass::CommandStreamReceiver::lastAdditionalKernelExecInfo;
     using BaseClass::CommandStreamReceiver::lastKernelExecutionType;
-    using BaseClass::CommandStreamReceiver::lastMediaSamplerConfig;
     using BaseClass::CommandStreamReceiver::lastMemoryCompressionState;
     using BaseClass::CommandStreamReceiver::lastPreemptionMode;
     using BaseClass::CommandStreamReceiver::lastSentL3Config;

shared/test/common/mocks/mock_product_helper.cpp

@@ -434,7 +434,6 @@ struct UnknownProduct {
     };
 
     struct PipelineSelectStateSupport {
-        static constexpr bool mediaSamplerDopClockGate = false;
         static constexpr bool systolicMode = false;
     };
 

shared/test/unit_test/command_stream/command_stream_receiver_tests.cpp

@@ -350,7 +350,6 @@ HWTEST_F(CommandStreamReceiverTest, WhenCreatingCsrThenFlagsAreSetCorrectly) {
     EXPECT_TRUE(csr.stateComputeModeDirty);
     EXPECT_FALSE(csr.lastVmeSubslicesConfig);
     EXPECT_EQ(0u, csr.lastSentL3Config);
-    EXPECT_EQ(-1, csr.lastMediaSamplerConfig);
     EXPECT_EQ(PreemptionMode::Initial, csr.lastPreemptionMode);
     EXPECT_EQ(static_cast<uint32_t>(-1), csr.latestSentStatelessMocsConfig);
 }
@@ -2808,28 +2807,6 @@ HWTEST_F(CommandStreamReceiverTest, givenPipelineSelectStateNotInitedWhenTransit
 
     auto &commandStreamReceiver = pDevice->getUltCommandStreamReceiver<FamilyType>();
 
-    commandStreamReceiver.pipelineSupportFlags.systolicMode = false;
-    commandStreamReceiver.pipelineSupportFlags.mediaSamplerDopClockGate = true;
-
-    dispatchFlags.pipelineSelectArgs.mediaSamplerRequired = false;
-    commandStreamReceiver.handlePipelineSelectStateTransition(dispatchFlags);
-    EXPECT_TRUE(commandStreamReceiver.csrSizeRequestFlags.mediaSamplerConfigChanged);
-
-    commandStreamReceiver.pipelineSupportFlags.mediaSamplerDopClockGate = false;
-    commandStreamReceiver.lastMediaSamplerConfig = -1;
-    commandStreamReceiver.handlePipelineSelectStateTransition(dispatchFlags);
-    EXPECT_FALSE(commandStreamReceiver.csrSizeRequestFlags.mediaSamplerConfigChanged);
-
-    commandStreamReceiver.pipelineSupportFlags.mediaSamplerDopClockGate = true;
-    commandStreamReceiver.lastMediaSamplerConfig = 0;
-    commandStreamReceiver.handlePipelineSelectStateTransition(dispatchFlags);
-    EXPECT_FALSE(commandStreamReceiver.csrSizeRequestFlags.mediaSamplerConfigChanged);
-
-    dispatchFlags.pipelineSelectArgs.mediaSamplerRequired = true;
-    commandStreamReceiver.handlePipelineSelectStateTransition(dispatchFlags);
-    EXPECT_TRUE(commandStreamReceiver.csrSizeRequestFlags.mediaSamplerConfigChanged);
-
-    commandStreamReceiver.pipelineSupportFlags.mediaSamplerDopClockGate = false;
     commandStreamReceiver.pipelineSupportFlags.systolicMode = true;
 
     commandStreamReceiver.lastSystolicPipelineSelectMode = false;
@@ -2856,27 +2833,6 @@ HWTEST_F(CommandStreamReceiverTest,
 
     auto &commandStreamReceiver = pDevice->getUltCommandStreamReceiver<FamilyType>();
 
-    commandStreamReceiver.pipelineSupportFlags.systolicMode = false;
-    commandStreamReceiver.pipelineSupportFlags.mediaSamplerDopClockGate = true;
-
-    commandStreamReceiver.streamProperties.pipelineSelect.mediaSamplerDopClockGate.value = 1;
-    commandStreamReceiver.lastMediaSamplerConfig = -1;
-    dispatchFlags.pipelineSelectArgs.mediaSamplerRequired = false;
-    commandStreamReceiver.handlePipelineSelectStateTransition(dispatchFlags);
-    EXPECT_TRUE(commandStreamReceiver.csrSizeRequestFlags.mediaSamplerConfigChanged);
-
-    commandStreamReceiver.streamProperties.pipelineSelect.mediaSamplerDopClockGate.value = 0;
-    dispatchFlags.pipelineSelectArgs.mediaSamplerRequired = true;
-    commandStreamReceiver.handlePipelineSelectStateTransition(dispatchFlags);
-    EXPECT_TRUE(commandStreamReceiver.csrSizeRequestFlags.mediaSamplerConfigChanged);
-
-    commandStreamReceiver.streamProperties.pipelineSelect.mediaSamplerDopClockGate.value = 0;
-    commandStreamReceiver.lastMediaSamplerConfig = 1;
-    dispatchFlags.pipelineSelectArgs.mediaSamplerRequired = false;
-    commandStreamReceiver.handlePipelineSelectStateTransition(dispatchFlags);
-    EXPECT_FALSE(commandStreamReceiver.csrSizeRequestFlags.mediaSamplerConfigChanged);
-
-    commandStreamReceiver.pipelineSupportFlags.mediaSamplerDopClockGate = false;
     commandStreamReceiver.pipelineSupportFlags.systolicMode = true;
 
     commandStreamReceiver.streamProperties.pipelineSelect.systolicMode.value = 1;
@@ -5618,35 +5574,6 @@ HWTEST2_F(CommandStreamReceiverHwTest, givenSpecialPipelineSelectModeChangedWhen
     EXPECT_EQ(expectedSize, size);
 }
 
-HWTEST2_F(CommandStreamReceiverHwTest, givenCsrWhenPreambleSentThenRequiredCsrSizeDependsOnmediaSamplerConfigChanged, IsAtMostXeCore) {
-    using PIPELINE_SELECT = typename FamilyType::PIPELINE_SELECT;
-    using PIPE_CONTROL = typename FamilyType::PIPE_CONTROL;
-
-    auto &commandStreamReceiver = pDevice->getUltCommandStreamReceiver<FamilyType>();
-    CsrSizeRequestFlags csrSizeRequest = {};
-    DispatchFlags flags = DispatchFlagsHelper::createDefaultDispatchFlags();
-
-    commandStreamReceiver.isPreambleSent = true;
-
-    csrSizeRequest.mediaSamplerConfigChanged = false;
-    commandStreamReceiver.overrideCsrSizeReqFlags(csrSizeRequest);
-    auto mediaSamplerConfigNotChangedSize = commandStreamReceiver.getRequiredCmdStreamSize(flags, *pDevice);
-
-    csrSizeRequest.mediaSamplerConfigChanged = true;
-    commandStreamReceiver.overrideCsrSizeReqFlags(csrSizeRequest);
-    auto mediaSamplerConfigChangedSize = commandStreamReceiver.getRequiredCmdStreamSize(flags, *pDevice);
-
-    EXPECT_NE(mediaSamplerConfigChangedSize, mediaSamplerConfigNotChangedSize);
-    auto difference = mediaSamplerConfigChangedSize - mediaSamplerConfigNotChangedSize;
-
-    size_t expectedDifference = sizeof(PIPELINE_SELECT);
-    if (MemorySynchronizationCommands<FamilyType>::isBarrierPriorToPipelineSelectWaRequired(pDevice->getRootDeviceEnvironment())) {
-        expectedDifference += sizeof(PIPE_CONTROL);
-    }
-
-    EXPECT_EQ(expectedDifference, difference);
-}
-
 HWTEST_F(CommandStreamReceiverHwTest, givenPreambleSentWhenEstimatingFlushTaskSizeThenResultDependsOnAdditionalCmdsSize) {
     auto &commandStreamReceiver = pDevice->getUltCommandStreamReceiver<FamilyType>();
 

shared/test/unit_test/command_stream/stream_properties_tests.cpp

@@ -36,7 +36,6 @@ std::vector<StreamProperty *> getAllFrontEndProperties(FrontEndProperties &prope
 std::vector<StreamProperty *> getAllPipelineSelectProperties(PipelineSelectProperties &properties) {
     std::vector<StreamProperty *> allProperties;
     allProperties.push_back(&properties.modeSelected);
-    allProperties.push_back(&properties.mediaSamplerDopClockGate);
     allProperties.push_back(&properties.systolicMode);
     return allProperties;
 }

shared/test/unit_test/command_stream/stream_properties_tests_common.cpp

@@ -759,21 +759,14 @@ TEST(StreamPropertiesTests, whenSettingPipelineSelectPropertiesThenCorrectValueI
     productHelper.fillPipelineSelectPropertiesSupportStructure(pipelineSelectPropertiesSupport, *defaultHwInfo);
 
     for (auto modeSelected : ::testing::Bool()) {
-        for (auto mediaSamplerDopClockGate : ::testing::Bool()) {
+        for (auto systolicMode : ::testing::Bool()) {
-            for (auto systolicMode : ::testing::Bool()) {
+            properties.pipelineSelect.setPropertiesAll(modeSelected, systolicMode);
-                properties.pipelineSelect.setPropertiesAll(modeSelected, mediaSamplerDopClockGate, systolicMode);
 
-                EXPECT_EQ(modeSelected, properties.pipelineSelect.modeSelected.value);
+            EXPECT_EQ(modeSelected, properties.pipelineSelect.modeSelected.value);
-                if (pipelineSelectPropertiesSupport.mediaSamplerDopClockGate) {
+            if (pipelineSelectPropertiesSupport.systolicMode) {
-                    EXPECT_EQ(mediaSamplerDopClockGate, properties.pipelineSelect.mediaSamplerDopClockGate.value);
+                EXPECT_EQ(systolicMode, properties.pipelineSelect.systolicMode.value);
-                } else {
+            } else {
-                    EXPECT_EQ(-1, properties.pipelineSelect.mediaSamplerDopClockGate.value);
+                EXPECT_EQ(-1, properties.pipelineSelect.systolicMode.value);
-                }
-                if (pipelineSelectPropertiesSupport.systolicMode) {
-                    EXPECT_EQ(systolicMode, properties.pipelineSelect.systolicMode.value);
-                } else {
-                    EXPECT_EQ(-1, properties.pipelineSelect.systolicMode.value);
-                }
             }
         }
     }
@@ -782,17 +775,16 @@ TEST(StreamPropertiesTests, whenSettingPipelineSelectPropertiesThenCorrectValueI
 TEST(StreamPropertiesTests, givenModeSelectPipelineSelectPropertyWhenSettingChangedPropertyAndCheckIfDirtyThenExpectDirtyState) {
     MockPipelineSelectProperties pipeProperties{};
     pipeProperties.propertiesSupportLoaded = true;
-    pipeProperties.pipelineSelectPropertiesSupport.mediaSamplerDopClockGate = true;
     pipeProperties.pipelineSelectPropertiesSupport.systolicMode = true;
 
     constexpr bool constState = false;
     bool changingState = false;
-    pipeProperties.setPropertiesAll(changingState, constState, constState);
+    pipeProperties.setPropertiesAll(changingState, constState);
 
     EXPECT_TRUE(pipeProperties.isDirty());
 
     changingState = !changingState;
-    pipeProperties.setPropertiesAll(changingState, constState, constState);
+    pipeProperties.setPropertiesAll(changingState, constState);
 
     EXPECT_TRUE(pipeProperties.isDirty());
 }
@@ -800,26 +792,21 @@ TEST(StreamPropertiesTests, givenModeSelectPipelineSelectPropertyWhenSettingChan
 TEST(StreamPropertiesTests, givenSetAllPipelineSelectPropertiesWhenResettingStateThenResetValuesAndDirtyKeepSupportFlagLoaded) {
     MockPipelineSelectProperties psProperties{};
     psProperties.propertiesSupportLoaded = true;
-    psProperties.pipelineSelectPropertiesSupport.mediaSamplerDopClockGate = true;
     psProperties.pipelineSelectPropertiesSupport.systolicMode = true;
 
     bool modeSelected = false;
-    bool mediaSamplerDopClockGate = false;
     bool systolicMode = true;
-    psProperties.setPropertiesAll(modeSelected, mediaSamplerDopClockGate, systolicMode);
+    psProperties.setPropertiesAll(modeSelected, systolicMode);
     EXPECT_TRUE(psProperties.isDirty());
     EXPECT_EQ(0, psProperties.modeSelected.value);
-    EXPECT_EQ(0, psProperties.mediaSamplerDopClockGate.value);
     EXPECT_EQ(1, psProperties.systolicMode.value);
 
     psProperties.resetState();
     EXPECT_FALSE(psProperties.isDirty());
     EXPECT_EQ(-1, psProperties.modeSelected.value);
-    EXPECT_EQ(-1, psProperties.mediaSamplerDopClockGate.value);
     EXPECT_EQ(-1, psProperties.systolicMode.value);
 
     EXPECT_TRUE(psProperties.propertiesSupportLoaded);
-    EXPECT_TRUE(psProperties.pipelineSelectPropertiesSupport.mediaSamplerDopClockGate);
     EXPECT_TRUE(psProperties.pipelineSelectPropertiesSupport.systolicMode);
 }
 
@@ -887,53 +874,29 @@ TEST(StreamPropertiesTests, givenModeSelectedMediaSamplerClockGatePipelineSelect
     bool clearDirtyState = false;
     MockPipelineSelectProperties pipeProperties{};
     pipeProperties.propertiesSupportLoaded = true;
-    pipeProperties.pipelineSelectPropertiesSupport.mediaSamplerDopClockGate = false;
 
     bool modeSelected = false;
-    bool mediaSamplerDopClockGate = false;
+    pipeProperties.setPropertiesModeSelected(modeSelected, clearDirtyState);
-    pipeProperties.setPropertiesModeSelectedMediaSamplerClockGate(modeSelected, mediaSamplerDopClockGate, clearDirtyState);
     EXPECT_TRUE(pipeProperties.isDirty());
     EXPECT_EQ(0, pipeProperties.modeSelected.value);
-    EXPECT_EQ(-1, pipeProperties.mediaSamplerDopClockGate.value);
 
-    pipeProperties.setPropertiesModeSelectedMediaSamplerClockGate(modeSelected, mediaSamplerDopClockGate, clearDirtyState);
+    pipeProperties.setPropertiesModeSelected(modeSelected, clearDirtyState);
     EXPECT_FALSE(pipeProperties.isDirty());
 
-    pipeProperties.pipelineSelectPropertiesSupport.mediaSamplerDopClockGate = true;
-    pipeProperties.setPropertiesModeSelectedMediaSamplerClockGate(modeSelected, mediaSamplerDopClockGate, clearDirtyState);
-    EXPECT_TRUE(pipeProperties.isDirty());
-    EXPECT_EQ(0, pipeProperties.modeSelected.value);
-    EXPECT_EQ(0, pipeProperties.mediaSamplerDopClockGate.value);
-
-    pipeProperties.setPropertiesModeSelectedMediaSamplerClockGate(modeSelected, mediaSamplerDopClockGate, clearDirtyState);
-    EXPECT_FALSE(pipeProperties.isDirty());
-    EXPECT_EQ(0, pipeProperties.modeSelected.value);
-    EXPECT_EQ(0, pipeProperties.mediaSamplerDopClockGate.value);
-
     modeSelected = true;
-    mediaSamplerDopClockGate = true;
+    pipeProperties.setPropertiesModeSelected(modeSelected, clearDirtyState);
-    pipeProperties.setPropertiesModeSelectedMediaSamplerClockGate(modeSelected, mediaSamplerDopClockGate, clearDirtyState);
     EXPECT_TRUE(pipeProperties.isDirty());
     EXPECT_EQ(1, pipeProperties.modeSelected.value);
-    EXPECT_EQ(1, pipeProperties.mediaSamplerDopClockGate.value);
 
-    pipeProperties.setPropertiesModeSelectedMediaSamplerClockGate(modeSelected, mediaSamplerDopClockGate, clearDirtyState);
+    pipeProperties.setPropertiesModeSelected(modeSelected, clearDirtyState);
     EXPECT_FALSE(pipeProperties.isDirty());
     EXPECT_EQ(1, pipeProperties.modeSelected.value);
-    EXPECT_EQ(1, pipeProperties.mediaSamplerDopClockGate.value);
-
-    pipeProperties.setPropertiesModeSelectedMediaSamplerClockGate(modeSelected, mediaSamplerDopClockGate, clearDirtyState);
-    EXPECT_FALSE(pipeProperties.isDirty());
-    EXPECT_EQ(1, pipeProperties.modeSelected.value);
-    EXPECT_EQ(1, pipeProperties.mediaSamplerDopClockGate.value);
 
     clearDirtyState = true;
     modeSelected = false;
-    mediaSamplerDopClockGate = false;
+    pipeProperties.setPropertiesModeSelected(modeSelected, clearDirtyState);
-    pipeProperties.setPropertiesModeSelectedMediaSamplerClockGate(modeSelected, mediaSamplerDopClockGate, clearDirtyState);
     EXPECT_FALSE(pipeProperties.isDirty());
     EXPECT_EQ(0, pipeProperties.modeSelected.value);
-    EXPECT_EQ(0, pipeProperties.mediaSamplerDopClockGate.value);
 }
 
 TEST(StreamPropertiesTests, givenStateBaseAddressSupportFlagStateWhenSettingPropertyAndCheckIfDirtyThenExpectCleanStateForNotSupportedAndDirtyForSupported) {
@@ -1491,9 +1454,8 @@ TEST(StreamPropertiesTests, givenAllStreamPropertiesSetWhenAllStreamPropertiesRe
     globalStreamProperties.frontEndState.setPropertiesAll(isCooperativeKernel, disableEuFusion, disableOverdispatch);
 
     bool modeSelected = false;
-    bool mediaSamplerDopClockGate = false;
     bool systolicMode = true;
-    globalStreamProperties.pipelineSelect.setPropertiesAll(modeSelected, mediaSamplerDopClockGate, systolicMode);
+    globalStreamProperties.pipelineSelect.setPropertiesAll(modeSelected, systolicMode);
 
     int32_t statelessMocs = 1;
     int64_t bindingTablePoolBaseAddress = 2;
@@ -1523,7 +1485,6 @@ TEST(StreamPropertiesTests, givenAllStreamPropertiesSetWhenAllStreamPropertiesRe
     EXPECT_EQ(-1, globalStreamProperties.frontEndState.singleSliceDispatchCcsMode.value);
 
     EXPECT_EQ(-1, globalStreamProperties.pipelineSelect.modeSelected.value);
-    EXPECT_EQ(-1, globalStreamProperties.pipelineSelect.mediaSamplerDopClockGate.value);
     EXPECT_EQ(-1, globalStreamProperties.pipelineSelect.systolicMode.value);
 
     EXPECT_EQ(-1, globalStreamProperties.stateBaseAddress.statelessMocs.value);

shared/test/unit_test/gen12lp/adln/test_product_helper_adln.cpp

@@ -101,7 +101,6 @@ ADLNTEST_F(AdlnProductHelper, givenProductHelperWhenGetCommandsStreamPropertiesS
     EXPECT_FALSE(productHelper->getFrontEndPropertyDisableOverDispatchSupport());
     EXPECT_FALSE(productHelper->getFrontEndPropertySingleSliceDispatchCcsModeSupport());
 
-    EXPECT_TRUE(productHelper->getPipelineSelectPropertyMediaSamplerDopClockGateSupport());
     EXPECT_FALSE(productHelper->getPipelineSelectPropertySystolicModeSupport());
 }
 

shared/test/unit_test/gen12lp/adlp/command_stream_receiver_hw_tests_adlp.cpp

@@ -1,5 +1,5 @@
 /*
- * Copyright (C) 2022-2023 Intel Corporation
+ * Copyright (C) 2022-2025 Intel Corporation
  *
  * SPDX-License-Identifier: MIT
  *
@@ -24,10 +24,8 @@ ADLPTEST_F(Gen12LpCommandStreamReceiverHwTests, givenSystolicModeChangedWhenFlus
     StreamProperties &streamProperties = commandStreamReceiver.getStreamProperties();
 
     commandStreamReceiver.isPreambleSent = true;
-    commandStreamReceiver.lastMediaSamplerConfig = false;
 
     flushTaskFlags.pipelineSelectArgs.systolicPipelineSelectMode = true;
-    flushTaskFlags.pipelineSelectArgs.mediaSamplerRequired = false;
 
     commandStreamReceiver.flushTask(commandStream,
                                     0,

shared/test/unit_test/gen12lp/adlp/test_product_helper_adlp.cpp

@@ -108,7 +108,6 @@ ADLPTEST_F(AdlpProductHelper, givenProductHelperWhenGetCommandsStreamPropertiesS
     EXPECT_FALSE(productHelper->getFrontEndPropertyDisableOverDispatchSupport());
     EXPECT_FALSE(productHelper->getFrontEndPropertySingleSliceDispatchCcsModeSupport());
 
-    EXPECT_TRUE(productHelper->getPipelineSelectPropertyMediaSamplerDopClockGateSupport());
     EXPECT_TRUE(productHelper->getPipelineSelectPropertySystolicModeSupport());
 }
 

shared/test/unit_test/gen12lp/adls/test_product_helper_adls.cpp

@@ -103,7 +103,6 @@ ADLSTEST_F(AdlsProductHelper, givenProductHelperWhenGetCommandsStreamPropertiesS
     EXPECT_FALSE(productHelper->getFrontEndPropertyDisableOverDispatchSupport());
     EXPECT_FALSE(productHelper->getFrontEndPropertySingleSliceDispatchCcsModeSupport());
 
-    EXPECT_TRUE(productHelper->getPipelineSelectPropertyMediaSamplerDopClockGateSupport());
     EXPECT_FALSE(productHelper->getPipelineSelectPropertySystolicModeSupport());
 }
 

shared/test/unit_test/gen12lp/dg1/test_product_helper_dg1.cpp

@@ -157,6 +157,5 @@ DG1TEST_F(Dg1ProductHelper, givenProductHelperWhenGetCommandsStreamPropertiesSup
     EXPECT_FALSE(productHelper->getFrontEndPropertyDisableOverDispatchSupport());
     EXPECT_FALSE(productHelper->getFrontEndPropertySingleSliceDispatchCcsModeSupport());
 
-    EXPECT_TRUE(productHelper->getPipelineSelectPropertyMediaSamplerDopClockGateSupport());
     EXPECT_FALSE(productHelper->getPipelineSelectPropertySystolicModeSupport());
 }

shared/test/unit_test/gen12lp/rkl/test_product_helper_rkl.cpp

@@ -105,7 +105,6 @@ RKLTEST_F(RklProductHelper, givenProductHelperWhenGetCommandsStreamPropertiesSup
     EXPECT_FALSE(productHelper->getFrontEndPropertyDisableOverDispatchSupport());
     EXPECT_FALSE(productHelper->getFrontEndPropertySingleSliceDispatchCcsModeSupport());
 
-    EXPECT_TRUE(productHelper->getPipelineSelectPropertyMediaSamplerDopClockGateSupport());
     EXPECT_FALSE(productHelper->getPipelineSelectPropertySystolicModeSupport());
 }
 

shared/test/unit_test/gen12lp/tgllp/product_helper_tests_tgllp.cpp

@@ -193,7 +193,6 @@ TGLLPTEST_F(TgllpProductHelper, givenProductHelperWhenGetCommandsStreamPropertie
     EXPECT_FALSE(productHelper->getFrontEndPropertyDisableOverDispatchSupport());
     EXPECT_FALSE(productHelper->getFrontEndPropertySingleSliceDispatchCcsModeSupport());
 
-    EXPECT_TRUE(productHelper->getPipelineSelectPropertyMediaSamplerDopClockGateSupport());
     EXPECT_FALSE(productHelper->getPipelineSelectPropertySystolicModeSupport());
 }
 

shared/test/unit_test/os_interface/product_helper_tests.cpp

@@ -611,7 +611,6 @@ HWTEST_F(ProductHelperTest, WhenFillingPipelineSelectPropertiesSupportThenExpect
     PipelineSelectPropertiesSupport pipelineSelectPropertiesSupport = {};
 
     productHelper->fillPipelineSelectPropertiesSupportStructure(pipelineSelectPropertiesSupport, pInHwInfo);
-    EXPECT_EQ(productHelper->getPipelineSelectPropertyMediaSamplerDopClockGateSupport(), pipelineSelectPropertiesSupport.mediaSamplerDopClockGate);
     EXPECT_EQ(productHelper->isSystolicModeConfigurable(pInHwInfo), pipelineSelectPropertiesSupport.systolicMode);
 }
 

shared/test/unit_test/xe2_hpg_core/bmg/product_helper_tests_bmg.cpp

@@ -64,7 +64,6 @@ BMGTEST_F(BmgProductHelper, givenProductHelperWhenGetCommandsStreamPropertiesSup
     EXPECT_TRUE(productHelper->getFrontEndPropertyDisableOverDispatchSupport());
     EXPECT_TRUE(productHelper->getFrontEndPropertySingleSliceDispatchCcsModeSupport());
 
-    EXPECT_FALSE(productHelper->getPipelineSelectPropertyMediaSamplerDopClockGateSupport());
     EXPECT_FALSE(productHelper->getPipelineSelectPropertySystolicModeSupport());
 }
 

shared/test/unit_test/xe2_hpg_core/lnl/product_helper_tests_lnl.cpp

@@ -64,7 +64,6 @@ LNLTEST_F(LnlProductHelper, givenProductHelperWhenGetCommandsStreamPropertiesSup
     EXPECT_TRUE(productHelper->getFrontEndPropertyDisableOverDispatchSupport());
     EXPECT_TRUE(productHelper->getFrontEndPropertySingleSliceDispatchCcsModeSupport());
 
-    EXPECT_FALSE(productHelper->getPipelineSelectPropertyMediaSamplerDopClockGateSupport());
     EXPECT_FALSE(productHelper->getPipelineSelectPropertySystolicModeSupport());
 }
 

shared/test/unit_test/xe3_core/product_helper_tests_xe3_core.cpp

@@ -56,7 +56,6 @@ XE3_CORETEST_F(Xe3CoreProductHelper, givenProductHelperWhenGetCommandsStreamProp
     EXPECT_TRUE(productHelper->getFrontEndPropertyDisableOverDispatchSupport());
     EXPECT_TRUE(productHelper->getFrontEndPropertySingleSliceDispatchCcsModeSupport());
 
-    EXPECT_FALSE(productHelper->getPipelineSelectPropertyMediaSamplerDopClockGateSupport());
     EXPECT_FALSE(productHelper->getPipelineSelectPropertySystolicModeSupport());
 }
 

shared/test/unit_test/xe_hpc_core/pvc/test_product_helper_pvc.cpp

@@ -139,7 +139,6 @@ PVCTEST_F(PvcProductHelper, givenProductHelperWhenGetCommandsStreamPropertiesSup
     EXPECT_TRUE(productHelper->getFrontEndPropertyDisableOverDispatchSupport());
     EXPECT_TRUE(productHelper->getFrontEndPropertySingleSliceDispatchCcsModeSupport());
 
-    EXPECT_FALSE(productHelper->getPipelineSelectPropertyMediaSamplerDopClockGateSupport());
     EXPECT_TRUE(productHelper->getPipelineSelectPropertySystolicModeSupport());
 }
 

shared/test/unit_test/xe_hpg_core/dg2/test_product_helper_dg2.cpp

@@ -80,7 +80,6 @@ DG2TEST_F(Dg2ProductHelper, givenProductHelperWhenGetCommandsStreamPropertiesSup
     EXPECT_TRUE(productHelper->getFrontEndPropertyDisableOverDispatchSupport());
     EXPECT_TRUE(productHelper->getFrontEndPropertySingleSliceDispatchCcsModeSupport());
 
-    EXPECT_FALSE(productHelper->getPipelineSelectPropertyMediaSamplerDopClockGateSupport());
     EXPECT_TRUE(productHelper->getPipelineSelectPropertySystolicModeSupport());
 }
 

shared/test/unit_test/xe_hpg_core/os_agnostic_product_helper_xe_lpg_tests.cpp

@@ -232,7 +232,6 @@ HWTEST2_F(XeLpgProductHelperTests, givenProductHelperWhenGetCommandsStreamProper
     EXPECT_TRUE(productHelper->getFrontEndPropertyDisableOverDispatchSupport());
     EXPECT_TRUE(productHelper->getFrontEndPropertySingleSliceDispatchCcsModeSupport());
 
-    EXPECT_FALSE(productHelper->getPipelineSelectPropertyMediaSamplerDopClockGateSupport());
     EXPECT_TRUE(productHelper->getPipelineSelectPropertySystolicModeSupport());
 }