add device conversion for uint64 #1373

nicolemarsaglia · 2024-08-28T23:48:16Z

No description provided.

nicolemarsaglia · 2024-08-28T23:51:07Z

src/libs/ascent/CMakeLists.txt

@cyrush I'm not committed to having this be the file we use for the conversion. but I can't get it all to link. am I missing something else in the cmake?

This is the error I'm getting:

[ 85%] Linking CXX executable lulesh_ser cd /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/build/ascent-develop/examples/proxies/lulesh2.0.3 && /usr/tce/packages/cmake/cmake-3.23.1/bin/cmake -E cmake_link_script CMakeFiles/lulesh_ser.dir/link.txt --verbose=1 /usr/tce/packages/gcc/gcc-11.2.1/bin/g++ -O3 -DNDEBUG -L/usr/tce/packages/cuda/cuda-12.0.0/nvidia/lib64 -L/usr/tce/packages/spectrum-mpi/ibm/spectrum-mpi-2020.08.19/lib -pthread CMakeFiles/lulesh_ser.dir/lulesh.cc.o "CMakeFiles/lulesh_ser.dir/lulesh-comm.cc.o" "CMakeFiles/lulesh_ser.dir/lulesh-viz.cc.o" "CMakeFiles/lulesh_ser.dir/lulesh-util.cc.o" "CMakeFiles/lulesh_ser.dir/lulesh-init.cc.o" -o lulesh_ser -Wl,-rpath,/usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/build/ascent-develop/lib:/usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/vtk-m-v2.1.0/lib:/usr/tce/packages/spectrum-mpi/ibm/spectrum-mpi-rolling-release/lib:/usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/raja-v2024.02.1/lib:/usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/umpire-2024.02.1/lib64:/usr/WS1/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/mfem-4.6/lib:/usr/WS1/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/zlib-1.3/lib:/usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/camp-v2024.02.1/lib:/usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/conduit-v0.9.1/lib:/usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/hdf5-1.14.1-2/lib: ../../../lib/libascent.so ../../../lib/libascent_flow.so ../../../lib/librover.so ../../../lib/libvtkh_rendering.so ../../../lib/libvtkh_filters.so ../../../lib/libvtkh_compositing.so ../../../lib/libvtkh_core.so ../../../lib/libvtkh_utils.so /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/vtk-m-v2.1.0/lib/libvtkm_filter_connected_components-2.1.so.2.1.0 /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/vtk-m-v2.1.0/lib/libvtkm_filter_contour-2.1.so.2.1.0 /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/vtk-m-v2.1.0/lib/libvtkm_filter_vector_analysis-2.1.so.2.1.0 /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/vtk-m-v2.1.0/lib/libvtkm_filter_mesh_info-2.1.so.2.1.0 /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/vtk-m-v2.1.0/lib/libvtkm_filter_field_conversion-2.1.so.2.1.0 /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/vtk-m-v2.1.0/lib/libvtkm_filter_field_transform-2.1.so.2.1.0 /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/vtk-m-v2.1.0/lib/libvtkm_filter_flow-2.1.so.2.1.0 /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/vtk-m-v2.1.0/lib/libvtkm_filter_multi_block-2.1.so.2.1.0 /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/vtk-m-v2.1.0/lib/libvtkm_filter_resampling-2.1.so.2.1.0 /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/vtk-m-v2.1.0/lib/libvtkm_filter_density_estimate-2.1.so.2.1.0 /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/vtk-m-v2.1.0/lib/libvtkm_filter_geometry_refinement-2.1.so.2.1.0 /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/vtk-m-v2.1.0/lib/libvtkm_filter_scalar_topology-2.1.so.2.1.0 /usr/tce/packages/spectrum-mpi/ibm/spectrum-mpi-rolling-release/lib/libmpiprofilesupport.so /usr/tce/packages/spectrum-mpi/ibm/spectrum-mpi-rolling-release/lib/libmpi_ibm.so /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/vtk-m-v2.1.0/lib/libvtkm_filter_zfp-2.1.so.2.1.0 /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/vtk-m-v2.1.0/lib/libvtkm_rendering-2.1.so.2.1.0 /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/vtk-m-v2.1.0/lib/libvtkm_filter_entity_extraction-2.1.so.2.1.0 /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/vtk-m-v2.1.0/lib/libvtkm_filter_clean_grid-2.1.so.2.1.0 /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/vtk-m-v2.1.0/lib/libvtkm_filter_image_processing-2.1.so.2.1.0 /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/vtk-m-v2.1.0/lib/libvtkm_filter_core-2.1.so.2.1.0 /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/vtk-m-v2.1.0/lib/libvtkm_worklet-2.1.so.2.1.0 /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/vtk-m-v2.1.0/lib/libvtkm_io-2.1.so.2.1.0 /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/vtk-m-v2.1.0/lib/libvtkm_cont-2.1.so.2.1.0 /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/vtk-m-v2.1.0/lib/libvtkmdiympi.so.2.1.0 ../../../lib/libdray.so /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/raja-v2024.02.1/lib/libRAJA.so /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/umpire-2024.02.1/lib64/libumpire.so /usr/WS1/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/mfem-4.6/lib/libmfem.so -Wl,-rpath,/usr/WS1/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/conduit-v0.9.1/lib -L/usr/WS1/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/conduit-v0.9.1/lib -lconduit -lconduit_relay -lconduit_blueprint /usr/WS1/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/zlib-1.3/lib/libz.so /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/camp-v2024.02.1/lib/libcamp.so /usr/tce/packages/cuda/cuda-12.0.0/nvidia/lib64/libcudart_static.a /usr/lib64/librt.so ../../../lib/libapcomp.so /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/conduit-v0.9.1/lib/libconduit_relay.so -lpthread -lrt /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/hdf5-1.14.1-2/lib/libhdf5.so.310.1.0 -ldl /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/conduit-v0.9.1/lib/libconduit_blueprint.so /usr/workspace/nicolem/lassen/src/ascent_warpx/ascent/scripts/build_ascent/install/conduit-v0.9.1/lib/libconduit.so ../../../lib/libascent_lodepng.so /usr/tce/packages/gcc/gcc-11.2.1/rh/usr/bin/../libexec/gcc/ppc64le-redhat-linux/11/ld: ../../../lib/libascent.so: undefined reference to `ascent::castUint64ToFloat64(unsigned long const*, double*, int)' collect2: error: ld returned 1 exit status

I think the issue may be that we are missing the export declaration ASCENT_API.

Also, we need to use a macro for __host__ __device__ decoration (for when cuda + hip are not there)

We can use
ASCENT_EXEC, which is defined in ascent_execution_policies.hpp

ascent/src/libs/ascent/runtimes/expressions/ascent_execution_policies.hpp

Line 89 in e504a28

#define ASCENT_EXEC inline __host__ __device__

@cyrush Thank you! I will do this

nicolemarsaglia · 2024-08-28T23:52:04Z

src/libs/ascent/runtimes/ascent_vtkh_data_adapter.cpp

oof. sorry for all the print statements from debugging warpx. ignore those. important stuff is line 1680

cyrush · 2024-09-04T22:21:26Z

src/libs/ascent/runtimes/ascent_vtkh_device_utils.hpp

+//---------------------------------------------------------------------------//
+// Non-device decorators
+//---------------------------------------------------------------------------//
+static const std::string memory_space = "host";


We can't declare and define an instance in a header b/c that will cause multiple defs.
You can declare here and then init in cpp file.

cyrush · 2024-09-04T22:22:16Z

src/libs/ascent/runtimes/ascent_vtkh_device_utils.hpp

+// Lambda decorators
+//---------------------------------------------------------------------------//
+
+#if defined(__CUDACC__) && !defined(DEBUG_CPU_ONLY)


we should use ascent's CUDA ifdefs instead of the compiler defines

cyrush · 2024-09-04T22:34:50Z

src/libs/ascent/runtimes/ascent_vtkh_device_utils.cpp

+ASCENT_EXEC void 
+VTKHDeviceAdapter::castUint64ToFloat64(const uint64_t* input, double* output, int size) {
+    Array<double> int_to_double(input, size);
+    output = int_to_double.get_ptr(memory_space);


Array<double> int_to_double(input, size);

I don't think this strategy will work. Creating a double array that points to the integer input.

cyrush · 2024-09-05T15:40:25Z

@nicolemarsaglia
Since this is in vtk-h -- I am thinking maybe we should be using a vtk-m worklet after all?

Sorry for flip flopping on this, but seems like the right path. Can you look into what that will take?

nicolemarsaglia · 2024-09-05T15:43:20Z

@nicolemarsaglia Since this is in vtk-h -- I am thinking maybe we should be using a vtk-m worklet after all?

Sorry for flip flopping on this, but seems like the right path. Can you look into what that will take?

All good! I have been thinking the same thing. Wish me luck!!

This reverts commit 030b95b.

cyrush · 2024-09-17T15:31:10Z

src/libs/ascent/runtimes/ascent_vtkh_data_adapter.cpp

+            vtkm::cont::Invoker invoker;
+            vtkh::VTKmTypeCast worklet;
+
+            invoker(worklet,input_arr,vtkm_arr);


This is a nice easy to use solution!

cyrush · 2024-09-17T15:31:38Z

src/libs/vtkh/utils/vtkm_array_utils.hpp

@@ -12,5 +13,25 @@ GetVTKMPointer(vtkm::cont::ArrayHandle<T> &handle)
  return handle.WritePortal().GetArray();
 }

+class VTKmTypeCast : public vtkm::worklet::WorkletMapField
+{


Looks good!

cyrush · 2024-09-17T15:31:57Z

src/libs/ascent/runtimes/ascent_vtkh_data_adapter.cpp

@@ -1891,68 +1893,97 @@ VTKHDataAdapter::AddField(const std::string &field_name,
        //   Array: valueType=x storageType=N4vtkm4cont15StorageTagBasicE 27 values occupying 216 bytes [0 1 2 ... 24 25 26]
        //   TypeList: N4vtkm4ListIJfdEEE
        // ***********************************************************************
+        // ***********************************************************************
+        // NOTE: int32 should work as of sept 10 2024 
+        // ***********************************************************************


VTK-m 2.2, or later?

It should have always worked, and it was working when I uncommented it, but the errors previously encountered make me too nervous to commit to changing it from something that works to something that might work sometimes. So I just left a note and if the else catch-all stops working for int32s we can try changing it.

Got it, thanks for clarifying.

add device conversion. stuck on libascent.so error

a03227a

nicolemarsaglia commented Aug 28, 2024

View reviewed changes

nicolemarsaglia added 3 commits August 30, 2024 16:27

add different policies

b28f475

errors on errors

b04a30d

idk what haps here

70496c9

cyrush reviewed Sep 4, 2024

View reviewed changes

wrangle all the defs I need. consolidate to .hpp

6a693b8

nicolemarsaglia added 9 commits September 9, 2024 13:44

worklet start. lassen down so can't build/test

46290b9

things build with doubs, now unsigned long long.

d5a2c52

it did. shocked gagged gooped

09f92c5

copy should be off

6b5245b

things work, gonna try out the one line reinterpret_cast

6903048

Merge branch 'develop' into task/08_2024_uint64_gpu_conversion

e69253e

clean up, use reinterpret_cast

030b95b

Revert "clean up, use reinterpret_cast"

b55a2ce

This reverts commit 030b95b.

nvm reinterpret_cast does not work. but my worklet does. REDEMPTION

53ef3ed

nicolemarsaglia requested a review from cyrush September 12, 2024 15:15

cyrush approved these changes Sep 17, 2024

View reviewed changes

nicolemarsaglia merged commit af8498b into develop Sep 17, 2024
22 checks passed

nicolemarsaglia deleted the task/08_2024_uint64_gpu_conversion branch September 17, 2024 16:26

Provide feedback

Saved searches

Use saved searches to filter your results more quickly

add device conversion for uint64 #1373

add device conversion for uint64 #1373

nicolemarsaglia commented Aug 28, 2024

nicolemarsaglia Aug 28, 2024

cyrush Aug 29, 2024 •

edited

Loading

nicolemarsaglia Aug 29, 2024

nicolemarsaglia Aug 28, 2024

cyrush Sep 4, 2024

cyrush Sep 4, 2024

cyrush Sep 4, 2024

cyrush commented Sep 5, 2024

nicolemarsaglia commented Sep 5, 2024

cyrush Sep 17, 2024

cyrush Sep 17, 2024

cyrush Sep 17, 2024

nicolemarsaglia Sep 17, 2024

cyrush Sep 17, 2024

add device conversion for uint64 #1373

add device conversion for uint64 #1373

Conversation

nicolemarsaglia commented Aug 28, 2024

Choose a reason for hiding this comment

cyrush Aug 29, 2024 • edited Loading

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

cyrush commented Sep 5, 2024

nicolemarsaglia commented Sep 5, 2024

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

Choose a reason for hiding this comment

cyrush Aug 29, 2024 •

edited

Loading