From 61231bc66cd66f0420a5d68aca44fc1da90e72c4 Mon Sep 17 00:00:00 2001
From: Tanguy Pruvot <tanguy.pruvot@gmail.com>
Date: Sat, 11 Mar 2017 11:19:20 +0100
Subject: [PATCH] fix various memory leaks on algo switch

---
 bench.cpp                 | 8 ++++++++
 lyra2/cuda_lyra2Z.cu      | 7 +++++++
 lyra2/lyra2Z.cu           | 3 +++
 x11/veltor.cu             | 2 +-
 x15/cuda_x15_whirlpool.cu | 6 ++----
 5 files changed, 21 insertions(+), 5 deletions(-)

diff --git a/bench.cpp b/bench.cpp
index a89afd5..1412158 100644
--- a/bench.cpp
+++ b/bench.cpp
@@ -8,6 +8,7 @@
 
 #include "miner.h"
 #include "algos.h"
+#include <cuda_runtime.h>
 
 int bench_algo = -1;
 
@@ -114,6 +115,13 @@ bool bench_algo_switch_next(int thr_id)
 	if (algo == ALGO_MJOLLNIR) algo++; // same as heavy
 	if (algo == ALGO_WHIRLCOIN) algo++; // same as whirlpool
 	if (algo == ALGO_WHIRLPOOLX) algo++; // disabled
+	// todo: algo switch from RPC 2.0
+	if (algo == ALGO_CRYPTOLIGHT) algo++;
+	if (algo == ALGO_CRYPTONIGHT) algo++;
+	if (algo == ALGO_WILDKECCAK) algo++;
+	if (algo == ALGO_JACKPOT) algo++; // to fix
+	if (algo == ALGO_QUARK) algo++; // to fix
+	if (algo == ALGO_LBRY && CUDART_VERSION < 7000) algo++;
 
 	if (device_sm[dev_id] && device_sm[dev_id] < 300) {
 		// incompatible SM 2.1 kernels...
diff --git a/lyra2/cuda_lyra2Z.cu b/lyra2/cuda_lyra2Z.cu
index 1dc215a..3978450 100644
--- a/lyra2/cuda_lyra2Z.cu
+++ b/lyra2/cuda_lyra2Z.cu
@@ -890,6 +890,13 @@ void lyra2Z_cpu_init_sm2(int thr_id, uint32_t threads)
 	cudaMallocHost(&h_GNonces[thr_id], 2 * sizeof(uint32_t));
 }
 
+__host__
+void lyra2Z_cpu_free(int thr_id)
+{
+	cudaFree(d_GNonces[thr_id]);
+	cudaFreeHost(h_GNonces[thr_id]);
+}
+
 __host__
 uint32_t lyra2Z_getSecNonce(int thr_id, int num)
 {
diff --git a/lyra2/lyra2Z.cu b/lyra2/lyra2Z.cu
index 42799af..eb01d58 100644
--- a/lyra2/lyra2Z.cu
+++ b/lyra2/lyra2Z.cu
@@ -15,6 +15,7 @@ extern void blake256_cpu_setBlock_80(uint32_t *pdata);
 
 extern void lyra2Z_cpu_init(int thr_id, uint32_t threads, uint64_t *d_matrix);
 extern void lyra2Z_cpu_init_sm2(int thr_id, uint32_t threads);
+extern void lyra2Z_cpu_free(int thr_id);
 extern uint32_t lyra2Z_cpu_hash_32(int thr_id, uint32_t threads, uint32_t startNonce, uint64_t *d_outputHash, bool gtx750ti);
 
 extern void lyra2Z_setTarget(const void *ptarget);
@@ -158,6 +159,8 @@ extern "C" void free_lyra2Z(int thr_id)
 	cudaFree(d_hash[thr_id]);
 	if (device_sm[dev_id] >= 350)
 		cudaFree(d_matrix[thr_id]);
+	lyra2Z_cpu_free(thr_id);
+
 	init[thr_id] = false;
 
 	cudaDeviceSynchronize();
diff --git a/x11/veltor.cu b/x11/veltor.cu
index d422349..4c6f8cd 100644
--- a/x11/veltor.cu
+++ b/x11/veltor.cu
@@ -174,8 +174,8 @@ extern "C" void free_veltor(int thr_id)
 	cudaThreadSynchronize();
 
 	cudaFree(d_hash[thr_id]);
+	cudaFree(d_resNonce[thr_id]);
 
-	cuda_check_cpu_free(thr_id);
 	init[thr_id] = false;
 
 	cudaDeviceSynchronize();
diff --git a/x15/cuda_x15_whirlpool.cu b/x15/cuda_x15_whirlpool.cu
index 30b85ba..933aa91 100644
--- a/x15/cuda_x15_whirlpool.cu
+++ b/x15/cuda_x15_whirlpool.cu
@@ -320,10 +320,8 @@ void whirlpool512_setBlock_80(void *pdata, const void *ptarget)
 __host__
 extern void x15_whirlpool_cpu_free(int thr_id)
 {
-	cudaFree(InitVector_RC);
-	cudaFree(b0);
-	cudaFree(b7);
-	cudaFree(d_resNonce[thr_id]);
+	if (d_resNonce[thr_id])
+		cudaFree(d_resNonce[thr_id]);
 }
 
 __global__