Cuda/Seeding2/TestDeviceCuts.cu

0001 // This file is part of the ACTS project.
0002 //
0003 // Copyright (C) 2016 CERN for the benefit of the ACTS project
0004 //
0005 // This Source Code Form is subject to the terms of the Mozilla Public
0006 // License, v. 2.0. If a copy of the MPL was not distributed with this
0007 // file, You can obtain one at https://mozilla.org/MPL/2.0/.
0008
0009 // Local include(s).
0010 #include "TestDeviceCuts.hpp"
0011
0012 // CUDA include(s).
0013 #include <cuda_runtime.h>
0014
0015 // System include(s).
0016 #include <stdexcept>
0017
0018 /// Helper macro for checking the return values of CUDA function calls
0019 #define CUDA_CHECK(EXP)                                         \
0020   do {                                                          \
0021     cudaError_t errorCode = EXP;                                \
0022     if (errorCode != cudaSuccess) {                             \
0023       throw std::runtime_error(                                 \
0024           "Failed to set up the custom seed filter functions"); \
0025     }                                                           \
0026   } while (false)
0027
0028 /// Code mimicking @c TestHostCuts::seedWeight
0029 __device__ float testSeedWeight(const Acts::Cuda::Details::SpacePoint& bottom,
0030                                 const Acts::Cuda::Details::SpacePoint&,
0031                                 const Acts::Cuda::Details::SpacePoint& top) {
0032   float weight = 0;
0033   if (bottom.radius > 150) {
0034     weight = 400;
0035   }
0036   if (top.radius < 150) {
0037     weight = 200;
0038   }
0039   return weight;
0040 }
0041
0042 /// Code mimicking @c TestHostCuts::singleSeedCut
0043 __device__ bool testSingleSeedCut(float weight,
0044                                   const Acts::Cuda::Details::SpacePoint& bottom,
0045                                   const Acts::Cuda::Details::SpacePoint&,
0046                                   const Acts::Cuda::Details::SpacePoint&) {
0047   return !(bottom.radius > 150. && weight < 380.);
0048 }
0049
0050 // Pointers to the device functions
0051 static __device__ Acts::Cuda::TripletFilterConfig::seedWeightFunc_t
0052     seedWeightDeviceFunc = testSeedWeight;
0053 static __device__ Acts::Cuda::TripletFilterConfig::singleSeedCutFunc_t
0054     singleSeedCutDeviceFunc = testSingleSeedCut;
0055
0056 Acts::Cuda::TripletFilterConfig testDeviceCuts() {
0057   // Create the filter configuration object.
0058   Acts::Cuda::TripletFilterConfig result;
0059
0060   // Set up the function pointer variables on it to point at the functions
0061   // implemented in this source file.
0062   CUDA_CHECK(cudaMemcpyFromSymbol(
0063       &(result.seedWeight), ::seedWeightDeviceFunc,
0064       sizeof(Acts::Cuda::TripletFilterConfig::seedWeightFunc_t)));
0065   CUDA_CHECK(cudaMemcpyFromSymbol(
0066       &(result.singleSeedCut), ::singleSeedCutDeviceFunc,
0067       sizeof(Acts::Cuda::TripletFilterConfig::singleSeedCutFunc_t)));
0068
0069   // Return the configured object.
0070   return result;
0071 }