BRL-CAD/embree
1
0
Fork 0
Code Issues Pull Requests Actions Packages Projects Releases Wiki Activity

rename memory types and account for host unified memory in buffer constructor

Browse Source
This commit is contained in:
Florian Reibold
2025-04-03 11:23:50 +02:00
committed by Daniel
parent 82fd9f914a
commit 9c7cb086a1
20 changed files with 48 additions and 46 deletions
Download Patch File Download Diff File Expand all files Collapse all files
common/sys/alloc.cpp
+5 -5
View File
@@ -44,7 +44,7 @@ namespace embree
assert((align & (align-1)) == 0);
void* ptr = nullptr;
if (mode == EMBREE_USM_SHARED_DEVICE_READ_ONLY)
if (mode == EmbreeUSMMode::DEVICE_READ_ONLY)
ptr = sycl::aligned_alloc_shared(align,size,*device,*context,sycl::ext::oneapi::property::usm::device_read_only());
else
ptr = sycl::aligned_alloc_shared(align,size,*device,*context);
@@ -66,16 +66,16 @@ namespace embree
assert((align & (align-1)) == 0);
void* ptr = nullptr;
if (type == EmbreeMemoryType::SHARED) {
if (mode == EMBREE_USM_SHARED_DEVICE_READ_ONLY)
if (type == EmbreeMemoryType::USM_SHARED) {
if (mode == EmbreeUSMMode::DEVICE_READ_ONLY)
ptr = sycl::aligned_alloc_shared(align,size,*device,*context,sycl::ext::oneapi::property::usm::device_read_only());
else
ptr = sycl::aligned_alloc_shared(align,size,*device,*context);
}
else if (type == EmbreeMemoryType::HOST) {
else if (type == EmbreeMemoryType::USM_HOST) {
ptr = sycl::aligned_alloc_host(align,size,*context);
}
else if (type == EmbreeMemoryType::DEVICE) {
else if (type == EmbreeMemoryType::USM_DEVICE) {
ptr = sycl::aligned_alloc_device(align,size,*device,*context);
}
else {
common/sys/alloc.h
+9 -9
View File
@@ -25,17 +25,17 @@ namespace embree
void alignedFree(void* ptr);
enum EmbreeUSMMode {
EMBREE_USM_SHARED = 0,
EMBREE_USM_SHARED_DEVICE_READ_WRITE = 0,
EMBREE_USM_SHARED_DEVICE_READ_ONLY = 1
enum class EmbreeUSMMode {
DEFAULT = 0,
DEVICE_READ_WRITE = 0,
DEVICE_READ_ONLY = 1
};
enum EmbreeMemoryType {
HOST = 0,
DEVICE = 1,
SHARED = 2,
UNKNOWN = 3
enum class EmbreeMemoryType {
USM_HOST = 0,
USM_DEVICE = 1,
USM_SHARED = 2,
MALLOC = 3
};
#if defined(EMBREE_SYCL_SUPPORT)
include/embree4/rtcore_geometry.isph
+1 -1
View File
@@ -185,7 +185,7 @@ RTC_API void* uniform rtcGetGeometryBufferData(RTCGeometry geometry, uniform RTC
/* Returns a pointer to the buffer data on the device. Returns the same pointer as
rtcGetGeometryBufferData if the device is no SYCL device or if Embree is executed on a
system with unified memory (e.g., iGPUs). */
RTC_API void* uniform rtcGetGeometryBufferDataDavice(RTCGeometry geometry, uniform RTCBufferType type, uniform unsigned int slot);
RTC_API void* uniform rtcGetGeometryBufferDataDevice(RTCGeometry geometry, uniform RTCBufferType type, uniform unsigned int slot);
/* Updates a geometry buffer. */
RTC_API void rtcUpdateGeometryBuffer(RTCGeometry geometry, uniform RTCBufferType type, uniform unsigned int slot);
kernels/common/buffer.h
+11 -5
View File
@@ -40,7 +40,7 @@ namespace embree
{
device->refInc();
ptr = alloc(ptr_in, shared, EmbreeMemoryType::SHARED);
ptr = alloc(ptr_in, shared, EmbreeMemoryType::USM_SHARED);
#if defined(EMBREE_SYCL_SUPPORT)
dshared = true;
dptr = ptr;
@@ -53,15 +53,21 @@ namespace embree
device->refInc();
#if defined(EMBREE_SYCL_SUPPORT)
if (device->is_gpu())
if (device->is_gpu() && !device->has_unified_memory())
{
ptr = alloc( ptr_in, shared, EmbreeMemoryType::UNKNOWN);
dptr = alloc(dptr_in, dshared, EmbreeMemoryType::DEVICE);
ptr = alloc( ptr_in, shared, EmbreeMemoryType::MALLOC);
dptr = alloc(dptr_in, dshared, EmbreeMemoryType::USM_DEVICE);
}
else if (device->is_gpu() && device->has_unified_memory())
{
ptr = alloc(ptr_in, shared, EmbreeMemoryType::USM_SHARED);
dshared = true;
dptr = ptr;
}
else
#endif
{
ptr = alloc(ptr_in, shared, EmbreeMemoryType::SHARED);
ptr = alloc(ptr_in, shared, EmbreeMemoryType::MALLOC);
#if defined(EMBREE_SYCL_SUPPORT)
dshared = true;
dptr = ptr;
kernels/common/device.cpp
+2 -2
View File
@@ -726,11 +726,11 @@ namespace embree
}
void* DeviceGPU::malloc(size_t size, size_t align) {
return alignedSYCLMalloc(&gpu_context,&gpu_device,size,align,EMBREE_USM_SHARED_DEVICE_READ_ONLY);
return alignedSYCLMalloc(&gpu_context,&gpu_device,size,align,EmbreeUSMMode::DEVICE_READ_ONLY);
}
void* DeviceGPU::malloc(size_t size, size_t align, EmbreeMemoryType type) {
return alignedSYCLMalloc(&gpu_context,&gpu_device,size,align,EMBREE_USM_SHARED_DEVICE_READ_ONLY,type);
return alignedSYCLMalloc(&gpu_context,&gpu_device,size,align,EmbreeUSMMode::DEVICE_READ_ONLY,type);
}
void DeviceGPU::free(void* ptr) {
kernels/common/device.h
+1 -1
View File
@@ -38,7 +38,7 @@ namespace embree
__forceinline pointer allocate( size_type n ) {
assert(device);
return (pointer) device->malloc(n*sizeof(T),alignment,EmbreeMemoryType::UNKNOWN);
return (pointer) device->malloc(n*sizeof(T),alignment,EmbreeMemoryType::MALLOC);
}
__forceinline void deallocate( pointer p, size_type n ) {
kernels/common/scene_instance.cpp
+2 -2
View File
@@ -17,7 +17,7 @@ namespace embree
gsubtype = GTY_SUBTYPE_INSTANCE_LINEAR;
world2local0 = one;
device->memoryMonitor(numTimeSteps*sizeof(AffineSpace3ff), false);
local2world = (AffineSpace3ff*) device->malloc(numTimeSteps*sizeof(AffineSpace3ff),16,EmbreeMemoryType::UNKNOWN);
local2world = (AffineSpace3ff*) device->malloc(numTimeSteps*sizeof(AffineSpace3ff),16,EmbreeMemoryType::MALLOC);
for (size_t i = 0; i < numTimeSteps; i++)
local2world[i] = one;
device->memoryMonitor(sizeof(*this), false);
@@ -37,7 +37,7 @@ namespace embree
return;
device->memoryMonitor(numTimeSteps_in*sizeof(AffineSpace3ff), false);
AffineSpace3ff* local2world2 = (AffineSpace3ff*) device->malloc(numTimeSteps_in*sizeof(AffineSpace3ff),16,EmbreeMemoryType::UNKNOWN);
AffineSpace3ff* local2world2 = (AffineSpace3ff*) device->malloc(numTimeSteps_in*sizeof(AffineSpace3ff),16,EmbreeMemoryType::MALLOC);
for (size_t i = 0; i < min(numTimeSteps, numTimeSteps_in); i++) {
local2world2[i] = local2world[i];
kernels/common/scene_instance_array.cpp
+1 -1
View File
@@ -59,7 +59,7 @@ namespace embree
numObjects = numScenes;
device->memoryMonitor(numObjects*sizeof(Accel*), false);
objects = (Accel**) device->malloc(numScenes*sizeof(Accel*),16,EmbreeMemoryType::UNKNOWN);
objects = (Accel**) device->malloc(numScenes*sizeof(Accel*),16,EmbreeMemoryType::MALLOC);
for (size_t i = 0; i < numObjects; ++i) {
Ref<Scene> scene = (Scene*) scenes[i];
objects[i] = scene.ptr;
kernels/rthwif/CMakeLists.txt
+1 -4
View File
@@ -176,7 +176,4 @@ IF (EMBREE_STATIC_LIB OR NOT EMBREE_RTHWIF_STATIC_LIB)
ENDIF()
ENDIF()
#ADD_SUBDIRECTORY(testing)
ADD_SUBDIRECTORY(testing)
kernels/sycl/rthwif_embree_builder.cpp
-1
View File
@@ -190,7 +190,6 @@ namespace embree
void* ptr = nullptr;
// TODO: modify memory monitor to account for host and device code separately
if (embree_device) embree_device->memoryMonitor(bytes,false);
ze_result_t result;
kernels/sycl/scene_sycl.cpp
+6 -6
View File
@@ -157,7 +157,7 @@ void Scene::syncWithDevice(sycl::queue queue, sycl::event* event)
if (geometries_device) {
device->free(geometries_device);
}
geometries_device = (Geometry**)device->malloc(sizeof(Geometry*) * geometries.size(), 16, EmbreeMemoryType::DEVICE);
geometries_device = (Geometry**)device->malloc(sizeof(Geometry*) * geometries.size(), 16, EmbreeMemoryType::USM_DEVICE);
}
if (num_geometries_changed || !dynamic_scene)
@@ -165,12 +165,12 @@ void Scene::syncWithDevice(sycl::queue queue, sycl::event* event)
if (offsets) {
device->free(offsets);
}
offsets = (size_t*)device->malloc(geometries.size() * sizeof(size_t), 16, EmbreeMemoryType::UNKNOWN);
offsets = (size_t*)device->malloc(geometries.size() * sizeof(size_t), 16, EmbreeMemoryType::MALLOC);
if (geometries_host) {
device->free(geometries_host);
}
geometries_host = (Geometry**)device->malloc(sizeof(Geometry*)*geometries.size(), 16, EmbreeMemoryType::UNKNOWN);
geometries_host = (Geometry**)device->malloc(sizeof(Geometry*)*geometries.size(), 16, EmbreeMemoryType::MALLOC);
}
size_t geometry_data_byte_size_ = 0;
@@ -189,7 +189,7 @@ void Scene::syncWithDevice(sycl::queue queue, sycl::event* event)
if (geometry_data_device) {
device->free(geometry_data_device);
}
geometry_data_device = (char*)device->malloc(geometry_data_byte_size, 16, EmbreeMemoryType::DEVICE);
geometry_data_device = (char*)device->malloc(geometry_data_byte_size, 16, EmbreeMemoryType::USM_DEVICE);
}
if (geometry_data_byte_size_changed || !dynamic_scene)
@@ -197,7 +197,7 @@ void Scene::syncWithDevice(sycl::queue queue, sycl::event* event)
if (geometry_data_host) {
device->free(geometry_data_host);
}
geometry_data_host = (char*)device->malloc(geometry_data_byte_size, 16, EmbreeMemoryType::UNKNOWN);
geometry_data_host = (char*)device->malloc(geometry_data_byte_size, 16, EmbreeMemoryType::MALLOC);
}
parallel_for(geometries.size(), [&] ( const size_t i ) {
@@ -222,7 +222,7 @@ void Scene::syncWithDevice(sycl::queue queue, sycl::event* event)
}
if (!scene_device) {
scene_device = (Scene*) device->malloc(sizeof(Scene), 16, EmbreeMemoryType::DEVICE);
scene_device = (Scene*) device->malloc(sizeof(Scene), 16, EmbreeMemoryType::USM_DEVICE);
}
sycl::event last_event = queue.memcpy(scene_device, (void*)this, sizeof(Scene));
tutorials/common/alloc/alloc.h
+1 -1
View File
@@ -28,7 +28,7 @@ namespace embree
void operator delete[](void* ptr) { alignedUSMFree(ptr); }
/*! aligned allocation using SYCL USM */
void* alignedUSMMalloc(size_t size, size_t align = 16, EmbreeUSMMode mode = EMBREE_USM_SHARED_DEVICE_READ_ONLY);
void* alignedUSMMalloc(size_t size, size_t align = 16, EmbreeUSMMode mode = EmbreeUSMMode::DEVICE_READ_ONLY);
void alignedUSMFree(void* ptr);
}
tutorials/common/tutorial/tutorial.cpp
+1 -1
View File
@@ -672,7 +672,7 @@ namespace embree
this->height = height;
if (pixels) alignedUSMFree(pixels);
pixels = (unsigned*) alignedUSMMalloc(width*height*sizeof(unsigned),64,EMBREE_USM_SHARED_DEVICE_READ_WRITE);
pixels = (unsigned*) alignedUSMMalloc(width*height*sizeof(unsigned),64,EmbreeUSMMode::DEVICE_READ_WRITE);
}
void TutorialApplication::set_scene (TutorialScene* in)
tutorials/hair_geometry/hair_geometry_device.cpp
+1 -1
View File
@@ -466,7 +466,7 @@ extern "C" void device_render (int* pixels,
{
/* create accumulator */
if (data.accu_width != width || data.accu_height != height) {
data.accu = (Vec3ff*) alignedUSMMalloc((width*height)*sizeof(Vec3ff),16,EMBREE_USM_SHARED_DEVICE_READ_WRITE);
data.accu = (Vec3ff*) alignedUSMMalloc((width*height)*sizeof(Vec3ff),16,EmbreeUSMMode::DEVICE_READ_WRITE);
data.accu_width = width;
data.accu_height = height;
for (unsigned int i=0; i<width*height; i++)
tutorials/ispc2cpp.sh
+1 -1
View File
@@ -59,7 +59,7 @@ sed -i.backup 's/foreach_tiled[ ]*([ ]*\([a-zA-Z0-9_]*\)[ ]*=[ ]*\([^ \.]*\)[ ]
sed -i.backup 's/foreach_unique[ ]*([ ]*\([[:alnum:]_]*\)[ ]*in[ ]*\([][[:alnum:]._]*\))/auto \1 = \2;/g' $2
sed -i.backup 's/new[ ]*\([a-zA-Z0-9_]*\)[ ]*\[\([^]]*\)\]; \/\/ EMBREE_USM_SHARED_DEVICE_READ_WRITE/(\1\*) alignedUSMMalloc((\2)\*sizeof(\1),16,EMBREE_USM_SHARED_DEVICE_READ_WRITE);/g' $2
sed -i.backup 's/new[ ]*\([a-zA-Z0-9_]*\)[ ]*\[\([^]]*\)\]; \/\/ EMBREE_USM_SHARED_DEVICE_READ_WRITE/(\1\*) alignedUSMMalloc((\2)\*sizeof(\1),16,EmbreeUSMMode::DEVICE_READ_WRITE);/g' $2
sed -i.backup 's/new[ ]*\([a-zA-Z0-9_]*\)[ ]*\[\([^]]*\)\]/(\1\*) alignedUSMMalloc((\2)\*sizeof(\1),16)/g' $2
sed -i.backup 's/delete[ ]*\[[ ]*\][ ]*\([a-zA-Z0-9_.\>\-]*\)/alignedUSMFree(\1)/g' $2
tutorials/motion_blur_geometry/motion_blur_geometry_device.cpp
+1 -1
View File
@@ -775,7 +775,7 @@ extern "C" void device_render (int* pixels,
/* create accumulator */
if (data.g_accu_width != width || data.g_accu_height != height) {
alignedUSMFree(data.g_accu);
data.g_accu = (Vec3ff*) alignedUSMMalloc((width*height)*sizeof(Vec3ff),16,EMBREE_USM_SHARED_DEVICE_READ_WRITE);
data.g_accu = (Vec3ff*) alignedUSMMalloc((width*height)*sizeof(Vec3ff),16,EmbreeUSMMode::DEVICE_READ_WRITE);
data.g_accu_width = width;
data.g_accu_height = height;
for (unsigned int i=0; i<width*height; i++)
tutorials/multi_instanced_geometry/multi_instanced_geometry_device.cpp
+1 -1
View File
@@ -355,7 +355,7 @@ extern "C" void device_render(int* pixels,
{
if (g_data.g_accu_width != width || g_data.g_accu_height != height) {
alignedUSMFree(g_data.g_accu);
g_data.g_accu = (Vec3ff*) alignedUSMMalloc((width*height)*sizeof(Vec3ff),16,EMBREE_USM_SHARED_DEVICE_READ_WRITE);
g_data.g_accu = (Vec3ff*) alignedUSMMalloc((width*height)*sizeof(Vec3ff),16,EmbreeUSMMode::DEVICE_READ_WRITE);
g_data.g_accu_width = width;
g_data.g_accu_height = height;
for (unsigned int i=0; i<width*height; i++)
tutorials/next_hit/next_hit_device.cpp
+1 -1
View File
@@ -542,7 +542,7 @@ extern "C" void device_render (unsigned* pixels,
if (!data.num_prev_hits || data.num_prev_hits_width != width || data.num_prev_hits_height != height)
{
alignedUSMFree(data.num_prev_hits);
data.num_prev_hits = (int*) alignedUSMMalloc(width*height*sizeof(int),16,EMBREE_USM_SHARED_DEVICE_READ_WRITE);
data.num_prev_hits = (int*) alignedUSMMalloc(width*height*sizeof(int),16,EmbreeUSMMode::DEVICE_READ_WRITE);
data.num_prev_hits_width = width;
data.num_prev_hits_height = height;
for (unsigned int i=0; i<width*height; i++)
tutorials/pathtracer/pathtracer_device.cpp
+1 -1
View File
@@ -1808,7 +1808,7 @@ extern "C" void device_render (int* pixels,
/* create accumulator */
if (data.accu_width != width || data.accu_height != height) {
alignedUSMFree(data.accu);
data.accu = (Vec3ff*) alignedUSMMalloc((width*height)*sizeof(Vec3ff),16,EMBREE_USM_SHARED_DEVICE_READ_WRITE);
data.accu = (Vec3ff*) alignedUSMMalloc((width*height)*sizeof(Vec3ff),16,EmbreeUSMMode::DEVICE_READ_WRITE);
data.accu_width = width;
data.accu_height = height;
for (unsigned int i=0; i<width*height; i++)
tutorials/quaternion_motion_blur/quaternion_motion_blur_device.cpp
+1 -1
View File
@@ -438,7 +438,7 @@ extern "C" void device_render (int* pixels,
if (data.g_accu_width != width || data.g_accu_height != height) {
alignedUSMFree(data.g_accu);
data.g_accu = (Vec3ff*) alignedUSMMalloc((width*height)*sizeof(Vec3ff),16,EMBREE_USM_SHARED_DEVICE_READ_WRITE);
data.g_accu = (Vec3ff*) alignedUSMMalloc((width*height)*sizeof(Vec3ff),16,EmbreeUSMMode::DEVICE_READ_WRITE);
data.g_accu_width = width;
data.g_accu_height = height;
for (unsigned int i=0; i<width*height; i++)