GPU: Plenty of clang-format fixes

Author: davidrohr

GPU/GPUTracking/Base/GPUParam.h

@@ -59,10 +59,10 @@ struct GPUParam_t {
   int32_t continuousMaxTimeBin;
   int32_t tpcCutTimeBin;
 
-  GPUTPCGeometry tpcGeometry;                       // TPC Geometry
-  GPUTPCGMPolynomialField polynomialField;          // Polynomial approx. of magnetic field for TPC GM
-  const uint32_t* occupancyMap;                     // Ptr to TPC occupancy map
-  uint32_t occupancyTotal;                          // Total occupancy in the TPC (nCl / nHbf)
+  GPUTPCGeometry tpcGeometry;              // TPC Geometry
+  GPUTPCGMPolynomialField polynomialField; // Polynomial approx. of magnetic field for TPC GM
+  const uint32_t* occupancyMap;            // Ptr to TPC occupancy map
+  uint32_t occupancyTotal;                 // Total occupancy in the TPC (nCl / nHbf)
 
   GPUParamSector SectorParam[GPUCA_NSECTORS];
 

GPU/GPUTracking/Base/GPUReconstruction.h

@@ -52,7 +52,7 @@ namespace gpu_reconstruction_kernels
 {
 struct deviceEvent;
 class threadContext;
-}
+} // namespace gpu_reconstruction_kernels
 
 class GPUReconstruction
 {
@@ -193,7 +193,7 @@ class GPUReconstruction
   bool IsInitialized() const { return mInitialized; }
   void SetSettings(float solenoidBzNominalGPU, const GPURecoStepConfiguration* workflow = nullptr);
   void SetSettings(const GPUSettingsGRP* grp, const GPUSettingsRec* rec = nullptr, const GPUSettingsProcessing* proc = nullptr, const GPURecoStepConfiguration* workflow = nullptr);
-  void SetResetTimers(bool reset) { mProcessingSettings.resetTimers = reset; } // May update also after Init()
+  void SetResetTimers(bool reset) { mProcessingSettings.resetTimers = reset; }     // May update also after Init()
   void SetDebugLevelTmp(int32_t level) { mProcessingSettings.debugLevel = level; } // Temporarily, before calling SetSettings()
   void UpdateSettings(const GPUSettingsGRP* g, const GPUSettingsProcessing* p = nullptr, const GPUSettingsRecDynamic* d = nullptr);
   void UpdateDynamicSettings(const GPUSettingsRecDynamic* d);

GPU/GPUTracking/Base/GPUReconstructionConvert.cxx

@@ -923,7 +923,7 @@ void zsEncoderDenseLinkBased::decodePage(std::vector<o2::tpc::Digit>& outputBuff
       if (decLinkX & 0b00100000) {
         bitmaskL2.set();
       } else {
-        bitmaskL2 = std::bitset<10>(((((uint16_t)decLinkX) & 0b11000000) << 2) | (uint16_t) * ((const uint8_t*)decPagePtr));
+        bitmaskL2 = std::bitset<10>(((((uint16_t)decLinkX) & 0b11000000) << 2) | (uint16_t)*((const uint8_t*)decPagePtr));
         decPagePtr += sizeof(uint8_t);
       }
 

GPU/GPUTracking/Base/GPUReconstructionProcessing.h

@@ -28,7 +28,7 @@ namespace gpu_reconstruction_kernels
 {
 struct deviceEvent {
   constexpr deviceEvent() = default;
-  constexpr deviceEvent(std::nullptr_t p) : v(nullptr){};
+  constexpr deviceEvent(std::nullptr_t p) : v(nullptr) {};
   template <class T>
   void set(T val)
   {

GPU/GPUTracking/Base/cuda/GPUReconstructionCUDA.cu

@@ -496,7 +496,7 @@ size_t GPUReconstructionCUDA::GPUMemCpy(void* dst, const void* src, size_t size,
     for (int32_t k = 0; k < nEvents; k++) {
       GPUChkErr(cudaStreamWaitEvent(mInternals->Streams[stream], evList[k].get<cudaEvent_t>(), 0));
     }
-    GPUChkErr(cudaMemcpyAsync(dst, src, size, toGPU == -2 ? cudaMemcpyDeviceToDevice : toGPU ? cudaMemcpyHostToDevice : cudaMemcpyDeviceToHost, mInternals->Streams[stream]));
+    GPUChkErr(cudaMemcpyAsync(dst, src, size, toGPU == -2 ? cudaMemcpyDeviceToDevice : (toGPU ? cudaMemcpyHostToDevice : cudaMemcpyDeviceToHost), mInternals->Streams[stream]));
   }
   if (ev) {
     GPUChkErr(cudaEventRecord(ev->get<cudaEvent_t>(), mInternals->Streams[stream == -1 ? 0 : stream]));

GPU/GPUTracking/DataTypes/GPUOutputControl.h

@@ -78,7 +78,7 @@ struct GPUTrackingOutputs {
   static constexpr size_t count() { return sizeof(GPUTrackingOutputs) / sizeof(GPUOutputControl); }
   GPUOutputControl* asArray() { return (GPUOutputControl*)this; }
   size_t getIndex(const GPUOutputControl& v) { return &v - (const GPUOutputControl*)this; }
-  static int32_t getIndex(GPUOutputControl GPUTrackingOutputs::*v) { return &(((GPUTrackingOutputs*)(0x10000))->*v) - (GPUOutputControl*)(0x10000); }
+  static int32_t getIndex(GPUOutputControl GPUTrackingOutputs::* v) { return &(((GPUTrackingOutputs*)(0x10000))->*v) - (GPUOutputControl*)(0x10000); }
 };
 
 } // namespace o2::gpu

GPU/GPUTracking/DataTypes/GPUSettings.h

@@ -73,9 +73,9 @@ struct GPUSettingsTF {
 
 // Settings defining the setup of the GPUReconstruction processing (basically selecting the device / class instance)
 struct GPUSettingsDeviceBackend {
-  uint32_t deviceType = GPUDataTypes::DeviceType::CPU;     // Device type, shall use GPUDataTypes::DEVICE_TYPE constants, e.g. CPU / CUDA
-  uint8_t forceDeviceType = 1;                             // Fail if device initialization fails, otherwise falls back to CPU
-  GPUReconstruction* master = nullptr;                     // GPUReconstruction master object
+  uint32_t deviceType = GPUDataTypes::DeviceType::CPU; // Device type, shall use GPUDataTypes::DEVICE_TYPE constants, e.g. CPU / CUDA
+  uint8_t forceDeviceType = 1;                         // Fail if device initialization fails, otherwise falls back to CPU
+  GPUReconstruction* master = nullptr;                 // GPUReconstruction master object
 };
 
 } // namespace o2::gpu

GPU/GPUTracking/DataTypes/GPUTRDTrack.h

@@ -107,9 +107,9 @@ class GPUTRDTrack_t : public T
   GPUd() void setHasPadrowCrossing() { mIsCrossingNeighbor |= (1U << 7); }
 
  protected:
-  float mChi2;                       // total chi2.
-  float mSignal{-1.f};               // electron Likelihood for track
-  uint32_t mRefGlobalTrackId;        // raw GlobalTrackID of the seeding track (either ITS-TPC or TPC)
+  float mChi2;                          // total chi2.
+  float mSignal{-1.f};                  // electron Likelihood for track
+  uint32_t mRefGlobalTrackId;           // raw GlobalTrackID of the seeding track (either ITS-TPC or TPC)
   int32_t mAttachedTracklets[kNLayers]; // indices of the tracklets attached to this track; -1 means no tracklet in that layer
   int16_t mCollisionId;                 // the collision ID of the tracklets attached to this track; is used to retrieve the BC information for this track after the tracking is done
   uint8_t mFlags;                       // bits 0 to 5 indicate whether track is findable in layer 0 to 5, bit 6 indicates an ambiguous track and bit 7 flags if the track is stopped in the TRD

GPU/GPUTracking/Global/GPUChain.h

@@ -46,7 +46,7 @@ class GPUChain
   virtual int32_t Finalize() = 0;
   virtual int32_t RunChain() = 0;
   virtual void MemorySize(size_t& gpuMem, size_t& pageLockedHostMem) = 0;
-  virtual void PrintMemoryStatistics(){};
+  virtual void PrintMemoryStatistics() {};
   virtual int32_t CheckErrorCodes(bool cpuOnly = false, bool forceShowErrors = false, std::vector<std::array<uint32_t, 4>>* fillErrors = nullptr) { return 0; }
   virtual bool SupportsDoublePipeline() { return false; }
   virtual int32_t FinalizePipelinedProcessing() { return 0; }

GPU/GPUTracking/Global/GPUChainTracking.h

@@ -43,7 +43,7 @@ class MatLayerCylSet;
 
 namespace o2::gpu
 {
-//class GPUTRDTrackerGPU;
+// class GPUTRDTrackerGPU;
 class GPUTPCGPUTracker;
 class GPUDisplayInterface;
 class GPUQA;