Add files via upload

network/spike_generation_cpu_cpp/PySpikeGenerationCPU.cpp

@@ -5,10 +5,15 @@
 
 namespace py = pybind11;
 
-PYBIND11_MODULE(PySpikeGenerationCPU, m) {
+PYBIND11_MODULE(PySpikeGenerationCPU, m)
+{
   m.doc() = "SpikeGenerationCPU Module";
   py::class_<SpikeGenerationCPU>(m, "SpikeGenerationCPU")
     .def(py::init<>())
+    .def("gpu_occupancy_export",
+      &SpikeGenerationCPU::gpu_occupancy_export)
+    .def("gpu_occupancy_import",
+      &SpikeGenerationCPU::gpu_occupancy_import)
     .def("spike_generation",
       &SpikeGenerationCPU::entrypoint);
 }

network/spike_generation_cpu_cpp/SpikeGenerationCPU.cpp

@@ -198,3 +198,23 @@ size_t SpikeGenerationCPU::lower_bound(float* data_ptr,
     }
     return start_of_range;
 };
+
+void SpikeGenerationCPU::gpu_occupancy_export(
+    size_t dim_x,
+    size_t dim_y,
+    size_t number_of_pattern,
+    size_t spike_dim,
+    int64_t setting_memory_addr,
+    size_t setting_dim_0,
+    size_t setting_dim_1)
+{
+    return;
+};
+
+void SpikeGenerationCPU::gpu_occupancy_import(
+    int64_t setting_memory_addr,
+    size_t setting_dim_0,
+    size_t setting_dim_1)
+{
+    return;
+};

network/spike_generation_cpu_cpp/SpikeGenerationCPU.h

@@ -6,33 +6,69 @@
 #include <cctype>
 #include <iostream>
 
-class SpikeGenerationCPU {
+class SpikeGenerationCPU
-public:
+{
+    public:
     SpikeGenerationCPU();
     ~SpikeGenerationCPU();
 
     void entrypoint(
-        int64_t input_pointer_addr, int64_t input_dim_0, int64_t input_dim_1,
+        int64_t input_pointer_addr,
-        int64_t input_dim_2, int64_t input_dim_3,
+        int64_t input_dim_0,
-        int64_t random_values_pointer_addr, int64_t random_values_dim_0,
+        int64_t input_dim_1,
-        int64_t random_values_dim_1, int64_t random_values_dim_2,
+        int64_t input_dim_2,
-        int64_t random_values_dim_3, int64_t output_pointer_addr,
+        int64_t input_dim_3,
-        int64_t output_dim_0, int64_t output_dim_1, int64_t output_dim_2,
+        int64_t random_values_pointer_addr,
-        int64_t output_dim_3, int64_t number_of_cpu_processes);
+        int64_t random_values_dim_0,
+        int64_t random_values_dim_1,
+        int64_t random_values_dim_2,
+        int64_t random_values_dim_3,
+        int64_t output_pointer_addr,
+        int64_t output_dim_0,
+        int64_t output_dim_1,
+        int64_t output_dim_2,
+        int64_t output_dim_3,
+        int64_t number_of_cpu_processes);
 
-private:
+    void gpu_occupancy_export(
-    void spike_generation(float* input_pointer, size_t input_dim_c0,
+        size_t dim_x,
-        size_t input_dim_c1, size_t input_dim_c2,
+        size_t dim_y,
+        size_t number_of_pattern,
+        size_t spike_dim,
+        int64_t setting_memory_addr,
+        size_t setting_dim_0,
+        size_t setting_dim_1);
+
+    void gpu_occupancy_import(
+        int64_t setting_memory_addr,
+        size_t setting_dim_0,
+        size_t setting_dim_1);
+
+    private:
+    void spike_generation(
+        float* input_pointer,
+        size_t input_dim_c0,
+        size_t input_dim_c1,
+        size_t input_dim_c2,
         float* random_values_pointer,
         size_t random_values_dim_c0,
         size_t random_values_dim_c1,
-        size_t random_values_dim_c2, int64_t* output_pointer,
+        size_t random_values_dim_c2,
-        size_t output_dim_c0, size_t output_dim_c1,
+        int64_t* output_pointer,
-        size_t output_dim_c2, size_t x_dim, size_t y_dim,
+        size_t output_dim_c0,
-        size_t spike_dim, size_t h_dim, size_t pattern_id);
+        size_t output_dim_c1,
+        size_t output_dim_c2,
+        size_t x_dim,
+        size_t y_dim,
+        size_t spike_dim,
+        size_t h_dim,
+        size_t pattern_id);
 
-    size_t lower_bound(float* data_ptr, size_t data_length,
+    size_t lower_bound(
-        size_t data_ptr_stride, float compare_to_value);
+        float* data_ptr,
+        size_t data_length,
+        size_t data_ptr_stride,
+        float compare_to_value);
 };
 
 #endif /* SPIKEGENERATIONCPU */

network/spike_generation_gpu_cpp_v1/SpikeGenerationGPU.cu

@@ -22,19 +22,19 @@ SpikeGenerationGPU::~SpikeGenerationGPU()
 void SpikeGenerationGPU::entrypoint(
     int64_t input_pointer_addr,
     int64_t input_dim_0,
-    int64_t input_dim_1, 
+    int64_t input_dim_1,
-    int64_t input_dim_2, 
+    int64_t input_dim_2,
     int64_t input_dim_3,
-    int64_t random_values_pointer_addr, 
+    int64_t random_values_pointer_addr,
     int64_t random_values_dim_0,
-    int64_t random_values_dim_1, 
+    int64_t random_values_dim_1,
     int64_t random_values_dim_2,
-    int64_t random_values_dim_3, 
+    int64_t random_values_dim_3,
     int64_t output_pointer_addr,
-    int64_t output_dim_0, 
+    int64_t output_dim_0,
-    int64_t output_dim_1, 
+    int64_t output_dim_1,
     int64_t output_dim_2,
-    int64_t output_dim_3, 
+    int64_t output_dim_3,
     int64_t number_of_cpu_processes)
 {
 
@@ -87,7 +87,7 @@ void SpikeGenerationGPU::entrypoint(
     size_t x_dim = output_dim_2;
     size_t y_dim = output_dim_2;
 
-    assert ((number_of_cpu_processes <= 0));
+    assert((number_of_cpu_processes <= 0));
 
     gpu_spike_generation(
         input_pointer,
@@ -208,9 +208,9 @@ void SpikeGenerationGPU::gpu_spike_generation(
     assert((x_dim < 65535));
     assert((y_dim < 65535));
 
-  // //////////////////////////////////////
+    // //////////////////////////////////////
-  // Calculate the distribution on the GPU
+    // Calculate the distribution on the GPU
-  // //////////////////////////////////////
+    // //////////////////////////////////////
 
     int min_grid_size;
     int block_size;
@@ -222,6 +222,15 @@ void SpikeGenerationGPU::gpu_spike_generation(
     status = cudaOccupancyMaxPotentialBlockSize(&min_grid_size, &block_size,
         (void*)kernel_spike_generation,
         dynamic_s_mem_size, max_threadable_tasks);
+    if (status != cudaSuccess)
+    {
+        std::cerr << "CUDA Runtime Error at: "
+            << __FILE__
+            << ":"
+            << __LINE__
+            << std::endl;
+        std::cerr << cudaGetErrorString(status) << std::endl;
+    }
     assert((status == cudaSuccess));
 
     grid_size = ((number_of_pattern * spike_dim) + block_size - 1) / block_size;
@@ -248,8 +257,17 @@ void SpikeGenerationGPU::gpu_spike_generation(
         h_dim,
         (number_of_pattern * spike_dim));
 
-    cudaDeviceSynchronize();
+    status = cudaDeviceSynchronize();
-
+    if (status != cudaSuccess)
+    {
+        std::cerr << "CUDA Runtime Error at: "
+            << __FILE__
+            << ":"
+            << __LINE__
+            << std::endl;
+        std::cerr << cudaGetErrorString(status) << std::endl;
+    }
+    assert((status == cudaSuccess));
     return;
 };
 

network/spike_generation_gpu_cpp_v2/kernel_spike_generation.cu

@@ -98,6 +98,15 @@ void occupancy_kernel_spike_generation(
     status = cudaOccupancyMaxPotentialBlockSize(
         &min_grid_size, &thread_block_size, (void*)kernel_spike_generation, 0,
         max_threadable_tasks);
+    if (status != cudaSuccess)
+    {
+        std::cerr << "CUDA Runtime Error at: "
+            << __FILE__
+            << ":"
+            << __LINE__
+            << std::endl;
+        std::cerr << cudaGetErrorString(status) << std::endl;
+    }
     assert((status == cudaSuccess));
 
     grid_size =