Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Jump to bottom

Issue with ManagedArray::Operator[] and RAJA #184

Open
rcarson3 opened this issue Jul 28, 2021 · 3 comments
Open

Issue with ManagedArray::Operator[] and RAJA #184

rcarson3 opened this issue Jul 28, 2021 · 3 comments

Comments

@rcarson3
Copy link
Member

@davidbeckingsale and others

I recently ran into a case for an application I'm working on where I had the initial chai::ExecutionSpace set to chai::ExecutionSpace::GPU for a chai::ManagedArray, and then I had it run within a RAJA forall loop on the CPU to initialize the data for some unit tests. Later on when trying to access the data on the host using the ManagedArray::Operator[] I would get a segfault. My understanding talking with @davidbeckingsale offline is that this should work as the data should now be allocated on the host/cpu.

I've included a MWE representing the issue.

Also, one other issue I noted was that doing a lambda capture by reference for the host RAJA loop also resulted in a segfault. I've included that below as well as commented out section of code. I will note that I've found this sort of lambda capture works fine if the original ExecutionSpace is set to the CPU.

Also, I'm using the following hashes for CHAI, RAJA, and Umpire:
CHAI: df3e8e0
RAJA: 3047fa720132d19ee143b1fcdacaa72971f5988c (v0.13.0 tagged release)
Umpire: 447f4640eff7b8f39d3c59404f3b03629b90c021 (v4.1.2 tagged release)

Additional information:
Compiled on rzansel with gcc/7.3.1, cuda/10.1.243, and cmake/3.14.5

#include "RAJA/RAJA.hpp"

#include "umpire/strategy/DynamicPool.hpp"
#include "umpire/Allocator.hpp"
#include "umpire/ResourceManager.hpp"

#include "chai/config.hpp"
#include "chai/ExecutionSpaces.hpp"
#include "chai/ManagedArray.hpp"

int main()
{

      auto& rm = umpire::ResourceManager::getInstance();
      auto host_allocator = rm.getAllocator("HOST");
#ifdef __CUDACC__
      auto device_allocator = rm.makeAllocator<umpire::strategy::DynamicPool>
                              ("DEVICE_pool", rm.getAllocator("DEVICE"));
#endif

      const int size = 5000;

      chai::ManagedArray<double> array(size, 
      std::initializer_list<chai::ExecutionSpace>{chai::CPU
#if defined(CHAI_ENABLE_CUDA) || defined(CHAI_ENABLE_HIP)
         , chai::GPU
#endif
         },
         std::initializer_list<umpire::Allocator>{host_allocator
#if defined(CHAI_ENABLE_CUDA) || defined(CHAI_ENABLE_HIP)
         , device_allocator
#endif
      },
      chai::ExecutionSpace::GPU);

      std::cout << "Running GPU runs" << std::endl;
      // This works
      RAJA::forall<RAJA::cuda_exec<256>>(RAJA::RangeSegment(0, size),
         [=] __device__ (int i) {
            array[i] = i;
      });

      std::cout << "Running CPU runs" << std::endl;
      // This should work but fails
      // RAJA::forall<RAJA::seq_exec>(RAJA::RangeSegment(0, size),
      //    [&] (int i) {
      //       array[i] = i;
      //    });
      // This works
      RAJA::forall<RAJA::seq_exec>(RAJA::RangeSegment(0, size),
         [=] (int i) {
            array[i] = i;
         });
      std::cout << "Printing out data" << std::endl;
      // These work
      // std::cout << array.data(chai::ExecutionSpace::CPU)[0] << std::endl;
      // std::cout << array.data()[0] << std::endl;
      // This should work since we last ran things on the CPU but fails
      std::cout << array[0] << std::endl;
      array.free();
      return 0;
}
@robinson96
Copy link
Contributor

robinson96 commented Jul 28, 2021 via email

@rcarson3
Copy link
Member Author

@robinson96 thanks for the quick response. I was unaware about the lambda capture by value requirement for the internal data transfer portion of things to work. The same goes about [] operator being undefined behavior outside of RAJA loops. I will definitely be careful about in my application codes and will need to update some libraries in regards to the above lambda capture issue.

If y'all would be open to it I can open a PR with some small documentations changes noting these requirements.

@robinson96
Copy link
Contributor

robinson96 commented Jul 28, 2021 via email

Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment
Assignees
No one assigned
Labels
None yet
Projects
None yet
Milestone
No milestone
Development

No branches or pull requests
2 participants
@robinson96 @rcarson3