[SYCL][HIP] Fix MemBufferFill for nvidia platform (#4629)

Author: AidanBeltonS

sycl/plugins/hip/pi_hip.cpp

@@ -3866,8 +3866,7 @@ pi_result hip_piEnqueueMemBufferFill(pi_queue command_queue, pi_mem buffer,
       result = retImplEv->start();
     }
 
-    auto dstDevice =
-        (uint8_t *)buffer->mem_.buffer_mem_.get_with_offset(offset);
+    auto dstDevice = buffer->mem_.buffer_mem_.get_with_offset(offset);
     auto stream = command_queue->get();
     auto N = size / pattern_size;
 
@@ -3915,7 +3914,8 @@ pi_result hip_piEnqueueMemBufferFill(pi_queue command_queue, pi_mem buffer,
         value = *(static_cast<const uint8_t *>(pattern) + step);
 
         // offset the pointer to the part of the buffer we want to write to
-        auto offset_ptr = dstDevice + (step * sizeof(uint8_t));
+        auto offset_ptr = reinterpret_cast<void *>(
+            reinterpret_cast<uint8_t *>(dstDevice) + (step * sizeof(uint8_t)));
 
         // set all of the pattern chunks
         result = PI_CHECK_ERROR(hipMemset2DAsync(

sycl/unittests/pi/EnqueueMemTest.cpp

@@ -74,12 +74,6 @@ class EnqueueMemTest : public testing::TestWithParam<detail::plugin> {
 
     detail::plugin plugin = GetParam();
 
-    if (plugin.getBackend() == sycl::backend::hip && sizeof(T) > 4) {
-      std::cerr << "HIP plugin doesn't support patterns larger than 4 bytes, "
-                   "skipping\n";
-      GTEST_SKIP();
-    }
-
     T inValues[_numElementsX] = {};
 
     for (size_t i = 0; i < _numElementsX; ++i) {