search

CHIP-SPV / chipStar

chipStar is a tool for compiling and running HIP/CUDA on SPIR-V via OpenCL or Level Zero APIs.
Other
182 stars 29 forks source link

hipStreamACb_StrmSyncTiming seems to cause GPU OOM for a rather small test #772

Closed pvelesko closed 5 months ago

pvelesko commented 6 months ago 
1062: CHIP error [TID 18703] [1707923182.868092708] : hipErrorNotInitialized (ZE_RESULT_ERROR_OUT_OF_DEVICE_MEMORY ) in /space/pvelesko/chipStar/dynamic-event-pools/src/backend/Level0/CHIPBackendLevel0.cc:1576:memCopyAsyncImpl
1062:

commenting out the kernel makes it go away

  HIPCHECK(hipMemcpy(A_d, A_h, Nbytes, hipMemcpyHostToDevice));

  const unsigned threadsPerBlock = 32;
  const unsigned blocks = (N_elmts + 31) / threadsPerBlock;

  hipLaunchKernelGGL((vector_square), dim3(blocks), dim3(threadsPerBlock), 0, 0, C_d, A_d,
                     N_elmts);
  HIPCHECK(hipDeviceSynchronize());
  HIPCHECK(hipMemcpy(C_h, C_d, Nbytes, hipMemcpyDeviceToHost));
  HIPCHECK(hipStreamAddCallback(0, Callback1, NULL, 0));
pvelesko commented 5 months ago

Fixed in #817