[SYCL] Overlapping subbuffers with different sizes need separate piMemBufferPartition (#10890)

Before this PR we were creating a single PI sub-buffer (of a smaller
size for the sycl/test-e2e/Basic/buffer/subbuffer_overlap.cpp added
here).

Suprisingly, on many platforms an "out-of-partition" access didn't cause
any issues and all the modified memory was copied between host/device.
However, that wasn't guaranteed by the PI calls SYCL RT was performing
and for some scenarios/HW the device-to-host copy after the second
kernel only brought back the smaller part of the original buffer
corresponding to the first smaller subbuffer.

Author: aelovikov-intel

sycl/source/detail/scheduler/graph_builder.cpp

@@ -654,8 +654,7 @@ AllocaCommandBase *Scheduler::GraphBuilder::findAllocaForReq(
       const Requirement *TmpReq = AllocaCmd->getRequirement();
       Res &= AllocaCmd->getType() == Command::CommandType::ALLOCA_SUB_BUF;
       Res &= TmpReq->MOffsetInBytes == Req->MOffsetInBytes;
-      Res &= TmpReq->MSYCLMemObj->getSizeInBytes() ==
-             Req->MSYCLMemObj->getSizeInBytes();
+      Res &= TmpReq->MAccessRange == Req->MAccessRange;
       Res &= AllowConst || !AllocaCmd->MIsConst;
     }
     return Res;

sycl/test-e2e/Basic/buffer/subbuffer_overlap.cpp

@@ -0,0 +1,44 @@
+// RUN: %{build} -o %t.out
+// RUN: %{run} %t.out
+// RUN: env SYCL_PI_TRACE=-1 %{run} %t.out 2>&1 | FileCheck %s
+
+#include <sycl/sycl.hpp>
+
+int main() {
+  sycl::queue q;
+  sycl::buffer<int, 1> b{1024};
+  sycl::id<1> start_offset{64};
+  size_t size = 16;
+  sycl::buffer<int, 1> sub1{b, start_offset, sycl::range<1>{size}};
+  sycl::buffer<int, 1> sub2{b, start_offset, sycl::range<1>{size * 2}};
+
+  int idx = 0;
+  for (auto &e : sycl::host_accessor{b})
+    e = idx++ % size;
+
+  // CHECK: piMemBufferPartition
+  // CHECK: pi_buffer_region origin/size : 256/64
+  q.submit([&](sycl::handler &cgh) {
+    sycl::accessor acc{sub1, cgh};
+    cgh.parallel_for(size, [=](auto id) { acc[id] += 1; });
+  });
+  // CHECK: piMemBufferPartition
+  // CHECK: pi_buffer_region origin/size : 256/128
+  q.submit([&](sycl::handler &cgh) {
+    sycl::accessor acc{sub2, cgh};
+    cgh.parallel_for(size * 2, [=](auto id) { acc[id] -= 1; });
+  });
+
+  // Print before asserts to ensure stream is flushed.
+  for (auto &e : sycl::host_accessor{sub2})
+    std::cout << e << " ";
+  std::cout << std::endl;
+
+  idx = 0;
+  for (auto &e : sycl::host_accessor{sub2}) {
+    assert(e == idx % size - idx / size);
+    ++idx;
+  }
+
+  return 0;
+}