ICudaTask.hpp

Go to the documentation of this file.

// NIST-developed software is provided by NIST as a public service. You may use, copy and distribute copies of the software in any medium, provided that you keep intact this entire notice. You may improve, modify and create derivative works of the software or any portion of the software, and you may copy and distribute such modifications or works. Modified works should carry a notice stating that you changed the software and should note the date and nature of any such change. Please explicitly acknowledge the National Institute of Standards and Technology as the source of the software.
// NIST-developed software is expressly provided "AS IS." NIST MAKES NO WARRANTY OF ANY KIND, EXPRESS, IMPLIED, IN FACT OR ARISING BY OPERATION OF LAW, INCLUDING, WITHOUT LIMITATION, THE IMPLIED WARRANTY OF MERCHANTABILITY, FITNESS FOR A PARTICULAR PURPOSE, NON-INFRINGEMENT AND DATA ACCURACY. NIST NEITHER REPRESENTS NOR WARRANTS THAT THE OPERATION OF THE SOFTWARE WILL BE UNINTERRUPTED OR ERROR-FREE, OR THAT ANY DEFECTS WILL BE CORRECTED. NIST DOES NOT WARRANT OR MAKE ANY REPRESENTATIONS REGARDING THE USE OF THE SOFTWARE OR THE RESULTS THEREOF, INCLUDING BUT NOT LIMITED TO THE CORRECTNESS, ACCURACY, RELIABILITY, OR USEFULNESS OF THE SOFTWARE.
// You are solely responsible for determining the appropriateness of using and distributing the software and you assume all risks associated with its use, including but not limited to the risks and costs of program errors, compliance with applicable laws, damage to or loss of data, programs or equipment, and the unavailability or interruption of operation. This software is not intended to be used in any situation where a failure could cause risk of injury or damage to property. The software developed by NIST employees is not subject to copyright protection within the United States.

#ifdef USE_CUDA
#ifndef HTGS_CUDATASK_HPP
#define HTGS_CUDATASK_HPP

#include <cuda_runtime_api.h>

#include <vector>
#include <unordered_map>
#include <algorithm>

#include <htgs/api/ITask.hpp>
namespace htgs {

template<class T>
class MemoryData;

template<class T, class U>
class ICudaTask : public ITask<T, U> {
  static_assert(std::is_base_of<IData, T>::value, "T must derive from IData");
  static_assert(std::is_base_of<IData, U>::value, "U must derive from IData");

 public:

  ICudaTask(int *cudaIds, size_t numGpus, bool autoEnablePeerAccess = true) {
    this->cudaIds = cudaIds;
    this->numGpus = numGpus;
  }


  virtual ~ICudaTask() override {}

  virtual void initializeCudaGPU() {}

  virtual void executeTask(std::shared_ptr<T> data) = 0;

  virtual void shutdownCuda() {}

  virtual std::string getName() override {
    return "Unnamed GPU ITask";
  }

  std::string getDotFillColor() override {
    return "green3";
  }

//  virtual std::string genDot(int flags, std::string dotId) override {
//    std::string inOutLabel = (((DOTGEN_FLAG_SHOW_IN_OUT_TYPES & flags) != 0) ? ("\nin: "+this->inTypeName()+"\nout: "+this->outTypeName()) : "");
//    std::string threadLabel = (((flags & DOTGEN_FLAG_SHOW_ALL_THREADING) != 0) ? "" : (" x" + std::to_string(this->getNumThreads())));
//    return dotId + "[label=\"" + this->getName()  + threadLabel + inOutLabel + "\",style=filled,fillcolor=forestgreen,shape=box,color=black,width=.2,height=.2];\n";
//  }

  virtual ITask <T, U> *copy() = 0;

  virtual void debug() override {}


  int getCudaId() {
    return this->cudaId;
  }

  bool requiresCopy(size_t pipelineId) {
    return std::find(this->nonPeerDevIds.begin(), this->nonPeerDevIds.end(),
      this->cudaIds[pipelineId]) != this->nonPeerDevIds.end();
  }

  template<class V>
  bool requiresCopy(std::shared_ptr<MemoryData<V>> data) {
    return this->requiresCopy(data->getPipelineId());
  }

  bool hasPeerToPeerCopy(size_t pipelineId) { return !requiresCopy((size_t)cudaId); }

  template<class V>
  bool autoCopy(V *destination, std::shared_ptr<MemoryData<V>> data, long numElems) {

    if (requiresCopy(data)) {
      cudaMemcpyPeerAsync((void *) destination,
                          this->cudaId,
                          (void *) data->get(),
                          this->cudaIds[data->getPipelineId()],
                          sizeof(V) * numElems,
                          this->stream);
      return true;
    } else {
      return false;
    }
  }

  void initialize() override final {
    this->cudaId = this->cudaIds[this->getPipelineId()];

    int numGpus;
    cudaGetDeviceCount(&numGpus);

    HTGS_ASSERT(this->cudaId < numGpus, "Error: Cuda ID: " << std::to_string(this->cudaId) << " is larger than the number of GPUs: " << std::to_string(numGpus));

    cudaSetDevice(this->cudaId);
    cudaStreamCreate(&stream);

    if (autoEnablePeerAccess) {

      for (size_t i = 0; i < this->numGpus; i++) {
        int peerId = this->cudaIds[i];
        if (peerId != this->cudaId) {
          int canAccess;
          cudaDeviceCanAccessPeer(&canAccess, this->cudaId, peerId);

          if (canAccess) {
            cudaDeviceEnablePeerAccess(peerId, 0);
          } else {
            this->nonPeerDevIds.push_back(peerId);
          }
        }
      }
    }

    this->initializeCudaGPU();
  }

  void shutdown() override final {
    this->shutdownCuda();
    cudaStreamDestroy(stream);
  }

  const cudaStream_t &getStream() const {
    return stream;
  }

  int *getCudaIds() {
    return this->cudaIds;
  }

  size_t getNumGPUs() {
    return this->numGpus;
  }

  void syncStream() {
    cudaStreamSynchronize(stream);
  }

 private:
  cudaStream_t stream; 
  int *cudaIds; 

  size_t numGpus; 
  int cudaId; 
  std::vector<int> nonPeerDevIds; 
  bool autoEnablePeerAccess; 
};

}
#endif //HTGS_CUDATASK_HPP
#endif //USE_CUDA