RPC API additions and compiler warning fixes

Makefile.am

@@ -54,7 +54,7 @@ cgminer_SOURCES += adl.c adl.h adl_functions.h
 cgminer_SOURCES += *.cl
 
 if HAS_SCRYPT
-cgminer_SOURCES += scrypt.c scrypt.h
+cgminer_SOURCES += scrypt.c scrypt.h scrypt-jane.c scrypt-jane.h
 endif
 
 endif

adl.c

@@ -99,19 +99,6 @@ static	ADL_OVERDRIVE5_POWERCONTROL_GET	ADL_Overdrive5_PowerControl_Get;
 static	ADL_OVERDRIVE5_POWERCONTROL_SET	ADL_Overdrive5_PowerControl_Set;
 static	ADL_OVERDRIVE5_FANSPEEDTODEFAULT_SET	ADL_Overdrive5_FanSpeedToDefault_Set;
 
-static	ADL_OVERDRIVE6_CAPABILITIES_GET ADL_Overdrive6_Capabilities_Get;
-static	ADL_OVERDRIVE6_FANSPEED_GET ADL_Overdrive6_FanSpeed_Get;
-static	ADL_OVERDRIVE6_THERMALCONTROLLER_CAPS ADL_Overdrive6_ThermalController_Caps;
-static	ADL_OVERDRIVE6_TEMPERATURE_GET ADL_Overdrive6_Temperature_Get;
-static	ADL_OVERDRIVE6_STATEINFO_GET ADL_Overdrive6_StateInfo_Get;
-static	ADL_OVERDRIVE6_CURRENTSTATUS_GET ADL_Overdrive6_CurrentStatus_Get;
-static	ADL_OVERDRIVE6_POWERCONTROL_CAPS ADL_Overdrive6_PowerControl_Caps;
-static	ADL_OVERDRIVE6_POWERCONTROLINFO_GET ADL_Overdrive6_PowerControlInfo_Get;
-static	ADL_OVERDRIVE6_POWERCONTROL_GET ADL_Overdrive6_PowerControl_Get;
-static	ADL_OVERDRIVE6_FANSPEED_SET ADL_Overdrive6_FanSpeed_Set;
-static	ADL_OVERDRIVE6_STATE_SET ADL_Overdrive6_State_Set;
-static	ADL_OVERDRIVE6_POWERCONTROL_SET ADL_Overdrive6_PowerControl_Set;
-
 #if defined (LINUX)
 	static void *hDLL;	// Handle to .so library
 #else
@@ -173,34 +160,6 @@ static bool init_overdrive5()
 	return true;
 }
 
-static bool init_overdrive6()
-{
-	ADL_Overdrive6_FanSpeed_Get = (ADL_OVERDRIVE6_FANSPEED_GET) GetProcAddress(hDLL,"ADL_Overdrive6_FanSpeed_Get");
-	ADL_Overdrive6_ThermalController_Caps = (ADL_OVERDRIVE6_THERMALCONTROLLER_CAPS)GetProcAddress (hDLL, "ADL_Overdrive6_ThermalController_Caps");
-	ADL_Overdrive6_Temperature_Get = (ADL_OVERDRIVE6_TEMPERATURE_GET)GetProcAddress (hDLL, "ADL_Overdrive6_Temperature_Get");
-	ADL_Overdrive6_Capabilities_Get = (ADL_OVERDRIVE6_CAPABILITIES_GET)GetProcAddress(hDLL, "ADL_Overdrive6_Capabilities_Get");
-	ADL_Overdrive6_StateInfo_Get = (ADL_OVERDRIVE6_STATEINFO_GET)GetProcAddress(hDLL, "ADL_Overdrive6_StateInfo_Get");
-	ADL_Overdrive6_CurrentStatus_Get = (ADL_OVERDRIVE6_CURRENTSTATUS_GET)GetProcAddress(hDLL, "ADL_Overdrive6_CurrentStatus_Get");
-	ADL_Overdrive6_PowerControl_Caps = (ADL_OVERDRIVE6_POWERCONTROL_CAPS)GetProcAddress(hDLL, "ADL_Overdrive6_PowerControl_Caps");
-	ADL_Overdrive6_PowerControlInfo_Get = (ADL_OVERDRIVE6_POWERCONTROLINFO_GET)GetProcAddress(hDLL, "ADL_Overdrive6_PowerControlInfo_Get");
-	ADL_Overdrive6_PowerControl_Get = (ADL_OVERDRIVE6_POWERCONTROL_GET)GetProcAddress(hDLL, "ADL_Overdrive6_PowerControl_Get");
-	ADL_Overdrive6_FanSpeed_Set  = (ADL_OVERDRIVE6_FANSPEED_SET)GetProcAddress(hDLL, "ADL_Overdrive6_FanSpeed_Set");
-	ADL_Overdrive6_State_Set = (ADL_OVERDRIVE6_STATE_SET)GetProcAddress(hDLL, "ADL_Overdrive6_State_Set");
-	ADL_Overdrive6_PowerControl_Set = (ADL_OVERDRIVE6_POWERCONTROL_SET) GetProcAddress(hDLL, "ADL_Overdrive6_PowerControl_Set");
-
-	if (!ADL_Overdrive6_FanSpeed_Get || !ADL_Overdrive6_ThermalController_Caps ||
-		!ADL_Overdrive6_Temperature_Get || !ADL_Overdrive6_Capabilities_Get ||
-		!ADL_Overdrive6_StateInfo_Get || !ADL_Overdrive6_CurrentStatus_Get ||
-		!ADL_Overdrive6_PowerControl_Caps || !ADL_Overdrive6_PowerControlInfo_Get ||
-		!ADL_Overdrive6_PowerControl_Get || !ADL_Overdrive6_FanSpeed_Set ||
-		!ADL_Overdrive6_State_Set || !ADL_Overdrive6_PowerControl_Set) {
-			applog(LOG_WARNING, "ATI ADL Overdrive6's API is missing");
-		return false;
-	}
-
-	return true;
-} 
-
 static bool prepare_adl(void)
 {
 	int result;

api.c

@@ -29,6 +29,10 @@
 #include "miner.h"
 #include "util.h"
 
+#ifdef HAVE_ADL
+#include "adl.h"
+#endif
+
 #if defined(USE_BFLSC) || defined(USE_AVALON) || defined(USE_HASHFAST) || defined(USE_BITFURY) || defined(USE_KLONDIKE) || defined(USE_KNC)
 #define HAVE_AN_ASIC 1
 #endif
@@ -2034,6 +2038,10 @@ static void gpustatus(struct io_data *io_data, int gpu, bool isjson, bool precom
 				(double)(cgpu->diff_rejected) / (double)(cgpu->diff1) : 0;
 		root = api_add_percent(root, "Device Rejected%", &rejp, false);
 		root = api_add_elapsed(root, "Device Elapsed", &(total_secs), true); // GPUs don't hotplug
+		double wu = cgpu->diff1 / total_secs * 60;
+		root = api_add_mhs(root, "Work Utility", &wu, false);
+		double wue = (cgpu->diff1 / total_secs * 60)/(cgpu->total_mhashes / total_secs * 100) / 10;
+		root = api_add_percent(root, "Work Utility Efficiency%", &wue, false);
 
 		root = print_data(root, buf, isjson, precom);
 		io_add(io_data, buf);

cgminer.c

@@ -58,6 +58,7 @@ char *curly = ":D";
 #include "driver-opencl.h"
 #include "bench_block.h"
 #include "scrypt.h"
+#include "scrypt-jane.h"
 #ifdef USE_USBUTILS
 #include "usbutils.h"
 #endif
@@ -113,6 +114,11 @@ static const bool opt_time = true;
 unsigned long long global_hashrate;
 unsigned long global_quota_gcd = 1;
 
+/* scrypt-jane, defaults suitable for YAC */
+unsigned int sj_minNf = 4;
+unsigned int sj_maxNf = 30;
+unsigned int sj_startTime = 1376184687;
+
 #if defined(HAVE_OPENCL) || defined(USE_USBUTILS)
 int nDevs;
 #endif
@@ -122,6 +128,7 @@ int opt_g_threads = -1;
 int gpu_threads;
 #ifdef USE_SCRYPT
 bool opt_scrypt;
+bool opt_scrypt_jane;
 #endif
 #endif
 bool opt_restart = true;
@@ -663,6 +670,11 @@ static char *set_int_0_to_10(const char *arg, int *i)
 	return set_int_range(arg, i, 0, 10);
 }
 
+static char *set_int_0_to_40(const char *arg, int *i)
+{
+	return set_int_range(arg, i, 0, 40);
+}
+
 #ifdef USE_AVALON
 static char *set_int_0_to_100(const char *arg, int *i)
 {
@@ -1445,6 +1457,18 @@ static struct opt_table opt_config_table[] = {
 	OPT_WITHOUT_ARG("--scrypt",
 			opt_set_bool, &opt_scrypt,
 			"Use the scrypt algorithm for mining"),
+	OPT_WITHOUT_ARG("--scrypt-jane",
+			opt_set_bool, &opt_scrypt_jane,
+			"Use the scrypt-jane algorithm for mining"),
+	OPT_WITH_ARG("--sj-nfmin",
+			set_int_0_to_40, opt_show_intval, &sj_minNf,
+			"Set min N factor for mining scrypt-jane"),
+	OPT_WITH_ARG("--sj-nfmax",
+			set_int_0_to_40, opt_show_intval, &sj_maxNf,
+			"Set max N factor for mining scrypt-jane"),
+	OPT_WITH_ARG("--sj-time",
+			opt_set_intval, opt_show_intval, &sj_startTime,
+			"Set StartTime for mining scrypt-jane"),
 	OPT_WITH_ARG("--cl-filename",
 		     set_cl_filename, NULL, NULL,
 		     "Sets the kernel .cl filename WITHOUT FILE EXTENSION - one value or comma separated list (e.g. scrypt130511_lantis,scrypt130511,scrypt130511_alexey)"),
@@ -3940,10 +3964,17 @@ static void regen_hash(struct work *work)
 
 static void rebuild_hash(struct work *work)
 {
-	if (opt_scrypt)
-		scrypt_regenhash(work);
-	else
+	if (opt_scrypt) {
+		if (opt_scrypt_jane) {
+			/* scrypt-jane */
+			sj_scrypt_regenhash(work);
+		} else {
+			/* standard scrypt */
+			scrypt_regenhash(work);
+		}
+	} else {
 		regen_hash(work);
+	}
 }
 
 static bool cnx_needed(struct pool *pool);
@@ -4348,12 +4379,13 @@ static bool test_work_current(struct work *work)
 		if (work->longpoll) {
 			work->work_block = ++work_block;
 			if (shared_strategy() || work->pool == current_pool()) {
-				if (opt_morenotices)
+				if (opt_morenotices) {
 					if (work->stratum) {
 						applog(LOG_NOTICE, "Stratum from %s requested work restart", pool->poolname);
 					} else {
 						applog(LOG_NOTICE, "%sLONGPOLL from %s requested work restart", work->gbt ? "GBT " : "", work->pool->poolname);
 					}
+				}
 				restart_threads();
 			}
 		}
@@ -4633,6 +4665,9 @@ void write_config(FILE *fcfg)
 					case KL_SCRYPT:
 						fprintf(fcfg, "scrypt");
 						break;
+					case KL_SCRYPT_JANE:
+						fprintf(fcfg, "scrypt-jane");
+						break;
 				}
 			}
 		};
@@ -4647,7 +4682,7 @@ void write_config(FILE *fcfg)
 			fprintf(fcfg, "%s%d", i > 0 ? "," : "", (int)gpus[i].opt_lg);
 		fputs("\",\n\"gpu-threads\" : \"", fcfg);
 		for(i = 0; i < nDevs; i++)
-			fprintf(fcfg, "%s%d", i > 0 ? "," : "", gpus[i].threads, gpus[i].threads);
+			fprintf(fcfg, "%s%d", i > 0 ? "," : "", gpus[i].threads);
 		fputs("\",\n\"thread-concurrency\" : \"", fcfg);
 		for(i = 0; i < nDevs; i++)
 			fprintf(fcfg, "%s%d", i > 0 ? "," : "", (int)gpus[i].opt_tc);
@@ -6376,7 +6411,11 @@ static void rebuild_nonce(struct work *work, uint32_t nonce)
 {
 	uint32_t *work_nonce = (uint32_t *)(work->data + 64 + 12);
 
-	*work_nonce = htole32(nonce);
+	if (opt_scrypt_jane) {
+		*work_nonce = htobe32(nonce);
+	} else {
+		*work_nonce = htole32(nonce);
+	}
 
 	rebuild_hash(work);
 }
@@ -8298,6 +8337,10 @@ int main(int argc, char *argv[])
 	if (!config_loaded)
 		load_default_config();
 
+	if (opt_scrypt_jane)
+		opt_set_bool(&opt_scrypt);
+
+
 	if (opt_benchmark) {
 		struct pool *pool;
 

configure.ac

@@ -481,6 +481,7 @@ AC_DEFINE_UNQUOTED([POCLBM_KERNNAME], ["poclbm130302"], [Filename for poclbm ker
 AC_DEFINE_UNQUOTED([DIAKGCN_KERNNAME], ["diakgcn121016"], [Filename for diakgcn kernel])
 AC_DEFINE_UNQUOTED([DIABLO_KERNNAME], ["diablo130302"], [Filename for diablo kernel])
 AC_DEFINE_UNQUOTED([SCRYPT_KERNNAME], ["scrypt130511"], [Filename for scrypt kernel])
+AC_DEFINE_UNQUOTED([SCRYPT_JANE_KERNNAME], ["scrypt-jane"], [Filename for scrypt-jane kernel])
 
 
 AC_SUBST(OPENCL_LIBS)

driver-opencl.c

@@ -34,6 +34,10 @@
 #include "adl.h"
 #include "util.h"
 
+#ifdef USE_SCRYPT
+#include "scrypt-jane.h"	/* sj_be32enc_vect */
+#endif
+
 /* TODO: cleanup externals ********************/
 
 #ifdef HAVE_CURSES
@@ -233,6 +237,8 @@ static enum cl_kernels select_kernel(char *arg)
 #ifdef USE_SCRYPT
 	if (!strcmp(arg, "scrypt"))
 		return KL_SCRYPT;
+	if (!strcmp(arg, "scrypt-jane"))
+		return KL_SCRYPT_JANE;
 #endif
 	return KL_NONE;
 }
@@ -1274,16 +1280,34 @@ static cl_int queue_diablo_kernel(_clState *clState, dev_blk_ctx *blk, cl_uint t
 }
 
 #ifdef USE_SCRYPT
+
 static cl_int queue_scrypt_kernel(_clState *clState, dev_blk_ctx *blk, __maybe_unused cl_uint threads)
 {
 	unsigned char *midstate = blk->work->midstate;
 	cl_kernel *kernel = &clState->kernel;
 	unsigned int num = 0;
 	cl_uint le_target;
 	cl_int status = 0;
+	uint32_t data[20];
+	unsigned int timestamp;
+	cl_uint nfactor;
+
+	if (opt_scrypt_jane) {
+		timestamp = bswap_32(*((unsigned int *)(blk->work->data + 17*4)));
+ 		nfactor = sj_GetNfactor(timestamp);
+		nfactor = (1 << (nfactor + 1));
+	}
 
 	le_target = *(cl_uint *)(blk->work->device_target + 28);
-	clState->cldata = blk->work->data;
+	if (!opt_scrypt_jane) {
+		clState->cldata = blk->work->data;
+	} else {
+		/* scrypt-jane */
+		applog(LOG_DEBUG, "Timestamp: %d, Nfactor: %d, Target: %x", timestamp, nfactor, le_target);
+		sj_be32enc_vect(data, (const uint32_t *)blk->work->data, 19);
+		clState->cldata = data;
+	}
+
 	status = clEnqueueWriteBuffer(clState->commandQueue, clState->CLbuffer0, true, 0, 80, clState->cldata, 0, NULL,NULL);
 
 	CL_SET_ARG(clState->CLbuffer0);
@@ -1292,6 +1316,8 @@ static cl_int queue_scrypt_kernel(_clState *clState, dev_blk_ctx *blk, __maybe_u
 	CL_SET_VARG(4, &midstate[0]);
 	CL_SET_VARG(4, &midstate[16]);
 	CL_SET_ARG(le_target);
+	if (opt_scrypt_jane)
+		CL_SET_ARG(nfactor);
 
 	return status;
 }
@@ -1600,6 +1626,9 @@ static bool opencl_thread_prepare(struct thr_info *thr)
 			case KL_SCRYPT:
 				cgpu->kname = "scrypt";
 				break;
+			case KL_SCRYPT_JANE:
+				cgpu->kname = "scrypt-jane";
+				break;
 #endif
 			case KL_POCLBM:
 				cgpu->kname = "poclbm";
@@ -1645,6 +1674,7 @@ static bool opencl_thread_init(struct thr_info *thr)
 			break;
 #ifdef USE_SCRYPT
 		case KL_SCRYPT:
+		case KL_SCRYPT_JANE:
 			thrdata->queue_kernel_parameters = &queue_scrypt_kernel;
 			break;
 #endif
@@ -1706,6 +1736,8 @@ static int64_t opencl_scanhash(struct thr_info *thr, struct work *work,
 	int64_t hashes;
 	int found = opt_scrypt ? SCRYPT_FOUND : FOUND;
 	int buffersize = opt_scrypt ? SCRYPT_BUFFERSIZE : BUFFERSIZE;
+	uint32_t *o;
+	uint32_t target;
 
 	/* Windows' timer resolution is only 15ms so oversample 5x */
 	if (gpu->dynamic && (++gpu->intervals * dynamic_us) > 70000) {
@@ -1739,6 +1771,9 @@ static int64_t opencl_scanhash(struct thr_info *thr, struct work *work,
 		size_t global_work_offset[1];
 
 		global_work_offset[0] = work->blk.nonce;
+		if (opt_scrypt_jane)
+			applog(LOG_DEBUG, "Nonce: %x, Global work size: %x, local work size: %x", work->blk.nonce, (unsigned)globalThreads[0], (unsigned)localThreads[0]);
+
 		status = clEnqueueNDRangeKernel(clState->commandQueue, *kernel, 1, global_work_offset,
 						globalThreads, localThreads, 0,  NULL, NULL);
 	} else
@@ -1756,6 +1791,12 @@ static int64_t opencl_scanhash(struct thr_info *thr, struct work *work,
 		return -1;
 	}
 
+	if (opt_scrypt_jane) {
+		o = thrdata->res;
+		target = *(uint32_t *)(work->target + 28);
+		applog(LOG_DEBUG, "Nonce: %x, Output buffer: %x %x %x %x %x %x %x %x Target: %x", work->blk.nonce, o[0], o[1], o[2], o[3], o[4], o[5], o[6], o[7], target);
+	}
+
 	/* The amount of work scanned can fluctuate when intensity changes
 	 * and since we do this one cycle behind, we increment the work more
 	 * than enough to prevent repeating work */

findnonce.c

@@ -18,6 +18,7 @@
 
 #include "findnonce.h"
 #include "scrypt.h"
+#include "scrypt-jane.h"
 
 const uint32_t SHA256_K[64] = {
 	0x428a2f98, 0x71374491, 0xb5c0fbcf, 0xe9b5dba5,

miner.h

@@ -406,6 +406,7 @@ enum cl_kernels {
 	KL_DIAKGCN,
 	KL_DIABLO,
 	KL_SCRYPT,
+	KL_SCRYPT_JANE,
 };
 
 enum dev_reason {
@@ -1095,6 +1096,11 @@ extern int opt_queue;
 extern int opt_scantime;
 extern int opt_expiry;
 
+/* scrypt-jane */
+extern unsigned int sj_minNf;
+extern unsigned int sj_maxNf;
+extern unsigned int sj_startTime;
+
 #ifdef USE_USBUTILS
 extern pthread_mutex_t cgusb_lock;
 extern pthread_mutex_t cgusbres_lock;
@@ -1187,6 +1193,7 @@ extern struct cgpu_info gpus[MAX_GPUDEVICES];
 extern int gpu_threads;
 #ifdef USE_SCRYPT
 extern bool opt_scrypt;
+extern bool opt_scrypt_jane;
 #else
 #define opt_scrypt (0)
 #endif
@@ -1631,4 +1638,6 @@ extern struct api_data *api_add_hs(struct api_data *root, char *name, double *da
 extern struct api_data *api_add_diff(struct api_data *root, char *name, double *data, bool copy_data);
 extern struct api_data *api_add_percent(struct api_data *root, char *name, double *data, bool copy_data);
 
+extern unsigned char sj_GetNfactor(int nTimestamp);
+
 #endif /* __MINER_H__ */