[CUDA] illegal memory access from cuda_piEnqueueMemBufferRead

[cgleggett]

I'm now seeing this error:

PI CUDA ERROR:
	Value:           700
	Name:            CUDA_ERROR_ILLEGAL_ADDRESS
	Description:     an illegal memory access was encountered
	Function:        cuda_piEnqueueMemBufferRead
	Source Location: /home/leggett/intel/git/llvm/sycl/plugins/cuda/pi_cuda.cpp:1881

the stack trace is:

#6  0x00007ffff6ea9edf in void cl::sycl::detail::plugin::checkPiResult<cl::sycl::runtime_error>(_pi_result) const [clone .constprop.163] [clone .cold.170] () from /opt/dpcpp/2020.04.03-cuda/lib/libsycl.so
#7  0x00007ffff7047588 in cl::sycl::detail::MemoryManager::map(cl::sycl::detail::SYCLMemObjI*, void*, std::shared_ptr<cl::sycl::detail::queue_impl>, cl::sycl::access::mode, unsigned int, cl::sycl::range<3>, cl::sycl::range<3>, cl::sycl::id<3>, unsigned int, std::vector<_pi_event*, std::allocator<_pi_event*> >, _pi_event*&) () from /opt/dpcpp/2020.04.03-cuda/lib/libsycl.so
#8  0x00007ffff708b381 in cl::sycl::detail::MapMemObject::enqueueImp() () from /opt/dpcpp/2020.04.03-cuda/lib/libsycl.so
#9  0x00007ffff7086a39 in cl::sycl::detail::Command::enqueue(cl::sycl::detail::EnqueueResultT&, cl::sycl::detail::BlockingT) () from /opt/dpcpp/2020.04.03-cuda/lib/libsycl.so
#10 0x00007ffff70955fe in cl::sycl::detail::Scheduler::GraphProcessor::enqueueCommand(cl::sycl::detail::Command*, cl::sycl::detail::EnqueueResultT&, cl::sycl::detail::BlockingT) () from /opt/dpcpp/2020.04.03-cuda/lib/libsycl.so
#11 0x00007ffff7090fd4 in cl::sycl::detail::Scheduler::addHostAccessor(cl::sycl::detail::AccessorImplHost*, bool) () from /opt/dpcpp/2020.04.03-cuda/lib/libsycl.so
#12 0x00007ffff6eb45de in cl::sycl::detail::addHostAccessorAndWait(cl::sycl::detail::AccessorImplHost*) () from /opt/dpcpp/2020.04.03-cuda/lib/libsycl.so
#13 0x00007ffff7083ea0 in cl::sycl::detail::stream_impl::flush() () from /opt/dpcpp/2020.04.03-cuda/lib/libsycl.so
#14 0x00007ffff708759a in cl::sycl::detail::ExecCGCommand::flushStreams() () from /opt/dpcpp/2020.04.03-cuda/lib/libsycl.so
#15 0x00007ffff7090de8 in cl::sycl::detail::Scheduler::addCG(std::unique_ptr<cl::sycl::detail::CG, std::default_delete<cl::sycl::detail::CG> >, std::shared_ptr<cl::sycl::detail::queue_impl>) () from /opt/dpcpp/2020.04.03-cuda/lib/libsycl.so
#16 0x00007ffff70c2a21 in cl::sycl::handler::finalize(cl::sycl::detail::code_location const&) () from /opt/dpcpp/2020.04.03-cuda/lib/libsycl.so
#17 0x00007ffff70c95ab in cl::sycl::detail::queue_impl::submit_impl(std::function<void (cl::sycl::handler&)> const&, std::shared_ptr<cl::sycl::detail::queue_impl>, cl::sycl::detail::code_location const&) () from /opt/dpcpp/2020.04.03-cuda/lib/libsycl.so
#18 0x00007ffff70cc030 in cl::sycl::queue::submit_impl(std::function<void (cl::sycl::handler&)>, cl::sycl::detail::code_location const&) () from /opt/dpcpp/2020.04.03-cuda/lib/libsycl.so
#19 0x00000000004049f5 in cl::sycl::queue::submit<find_prime_s(work*)::$_0>(find_prime_s(work*)::$_0, cl::sycl::detail::code_location const&) (this=0x10e5260, CGF=..., CodeLoc=...) at /opt/dpcpp/2020.04.03-cuda/bin/../include/sycl/CL/sycl/queue.hpp:183

this is with some test code that's run fine with all other previous versions of dpcpp. It fails with commit hash ee21d66 . The previous version that I tested it with was from 03/24

[jinz2014]

Could you please show the program for me to verify ?

[hiaselhans]

Did you try with SYCL_BE=PI_CUDA?

My guess is that this is due to USM relying on pi::useBackend(pi::Backend::SYCL_BE_PI_OPENCL) which basically quest's SYCL_BE env...

Maybe there is a way to get the backend from Context instead?

[cgleggett]

@hiaselhans : is SYCL_BE=PI_CUDA a compile line parameter? How do I specify it?

[hiaselhans]

it's an environment variable at run-time. like in GetStarted.

SYCL_BE=PI_CUDA ./my-executable

:)

[cgleggett]

@hiaselhans : setting the env var SYCL_BE=PI_CUDA did the trick.

[cgleggett]

BTW, I am compiling with -fsycl-targets=nvptx64-nvidia-cuda-sycldevice, and explicitly selecting the device with CUDA driver, which resolves as

   device.get_info<sycl::info::device::name>(): GeForce RTX 2080 SUPER
   device.get_info<cl::sycl::info::device::driver_version>():   CUDA 10.2
   platform.get_info<sycl::info::platform::name>(): NVIDIA CUDA
   platform.get_info<sycl::info::platform::version>():  CUDA 10.2

[hiaselhans]

@cgleggett i think this issue is resolved by #1490
can you try again?

[cgleggett]

@hiaselhans : looks good. thanks!

[hiaselhans]

@cgleggett perfect! So: i guess you can close this issue? :)