<html>
<head>
<meta http-equiv="Content-Type" content="text/html; charset=us-ascii">
<style type="text/css" style="display:none;"> P {margin-top:0;margin-bottom:0;} </style>
</head>
<body dir="ltr">
<div style="font-family: Calibri, Arial, Helvetica, sans-serif; font-size: 12pt; color: rgb(0, 0, 0);">
The pointer 'p' should be freed with 'omp_target_free` as in the 'omp_device_managed_memory.c' test. You're declaring a vector of vectors and only initializing the first dimension, the first '[]' operator will return a vector that hasn't been initialized, so
v[0][0] returns a nullptr.</div>
<div id="appendonsend"></div>
<hr style="display:inline-block;width:98%" tabindex="-1">
<div id="divRplyFwdMsg" dir="ltr"><font face="Calibri, sans-serif" style="font-size:11pt" color="#000000"><b>From:</b> Openmp-dev <openmp-dev-bounces@lists.llvm.org> on behalf of Itaru Kitayama via Openmp-dev <openmp-dev@lists.llvm.org><br>
<b>Sent:</b> Saturday, August 7, 2021 2:31 AM<br>
<b>To:</b> openmp-dev@lists.llvm.org <openmp-dev@lists.llvm.org><br>
<b>Subject:</b> [EXTERNAL] [Openmp-dev] Use of llvm_omp_target_alloc_shared API</font>
<div> </div>
</div>
<div class="BodyFragment"><font size="2"><span style="font-size:11pt;">
<div class="PlainText">Hi,<br>
Is this supposed to work? It does not on A100.<br>
<br>
#include <cstdlib><br>
#include <new><br>
#include <limits><br>
#include <iostream><br>
#include <vector><br>
#include <omp.h><br>
<br>
extern "C" {<br>
void* llvm_omp_target_alloc_shared(size_t, int);<br>
}<br>
<br>
template <class T><br>
struct Mallocator<br>
{<br>
typedef T value_type;<br>
<br>
Mallocator () = default;<br>
template <class U> constexpr Mallocator (const Mallocator <U>&) noexcept {}<br>
<br>
[[nodiscard]] T* allocate(std::size_t n) {<br>
if (n > std::numeric_limits<std::size_t>::max() / sizeof(T))<br>
throw std::bad_array_new_length();<br>
<br>
//if (auto p = static_cast<T*>(std::malloc(n*sizeof(T)))) {<br>
if (T* p = static_cast<T*>(llvm_omp_target_alloc_shared(n*sizeof(T), 0))) {<br>
report(p, n);<br>
return p;<br>
}<br>
<br>
throw std::bad_alloc();<br>
}<br>
<br>
void deallocate(T* p, std::size_t n) noexcept {<br>
report(p, n, 0);<br>
std::free(p);<br>
}<br>
<br>
private:<br>
void report(T* p, std::size_t n, bool alloc = true) const {<br>
std::cout << (alloc ? "Alloc: " : "Dealloc: ") << sizeof(T)*n<br>
<< " bytes at " << std::hex << std::showbase<br>
<< reinterpret_cast<void*>(p) << std::dec << '\n';<br>
}<br>
};<br>
<br>
template <class T, class U><br>
bool operator==(const Mallocator <T>&, const Mallocator <U>&) { return true; }<br>
template <class T, class U><br>
bool operator!=(const Mallocator <T>&, const Mallocator <U>&) { return false; }<br>
<br>
int main()<br>
{<br>
std::vector<std::vector<int, Mallocator<int> >,<br>
Mallocator<std::vector<int, Mallocator<int> > > > v1;<br>
v1.reserve(1024);<br>
<br>
auto p = v1.data();<br>
v1[0][0] = 999;<br>
v1[1][0] = 1;<br>
#pragma omp target parallel for device(0) is_device_ptr(p)<br>
for (int i=0;i < 2;i++) {<br>
printf("v1 %d %d\n", i, p[0][0]);<br>
}<br>
<br>
}<br>
_______________________________________________<br>
Openmp-dev mailing list<br>
Openmp-dev@lists.llvm.org<br>
<a href="https://lists.llvm.org/cgi-bin/mailman/listinfo/openmp-dev">https://lists.llvm.org/cgi-bin/mailman/listinfo/openmp-dev</a><br>
</div>
</span></font></div>
</body>
</html>