diff --git a/src/mame/drivers/saturn.c b/src/mame/drivers/saturn.c
index 9de1b671d6a..a74b8a75c6c 100644
--- a/src/mame/drivers/saturn.c
+++ b/src/mame/drivers/saturn.c
@@ -563,7 +563,8 @@ static WRITE32_HANDLER( saturn_scu_w )
 		case 0x08/4: case 0x28/4: case 0x48/4:  state->m_scu.size[DMA_CH] = ((state->m_scu_regs[offset] & ((offset == 2) ? 0x000fffff : 0xfff)) >> 0); break;
 		case 0x0c/4: case 0x2c/4: case 0x4c/4:
 			state->m_scu.src_add[DMA_CH] = (state->m_scu_regs[offset] & 0x100) ? 4 : 0;
-			state->m_scu.dst_add[DMA_CH] = 2 << (state->m_scu_regs[offset] & 7);
+			state->m_scu.dst_add[DMA_CH] = 1 << (state->m_scu_regs[offset] & 7);
+			if(state->m_scu.dst_add[DMA_CH] == 1) { state->m_scu.dst_add[DMA_CH] = 0; }
 			break;
 		case 0x10/4: case 0x30/4: case 0x50/4:
 			state->m_scu.enable_mask[DMA_CH] = (data & 0x100) >> 8;
@@ -661,14 +662,32 @@ static TIMER_CALLBACK( dma_lv2_ended )
 	DnMV_0(2);
 }
 
+static void scu_single_transfer(address_space *space, UINT32 src, UINT32 dst,UINT8 *src_shift)
+{
+	UINT32 src_data;
+
+	if(src & 1)
+	{
+		/* Road Blaster does a work ram h to color ram with offsetted source address, do some data rotation */
+		src_data = ((space->read_dword(src & 0x07fffffc) & 0x00ffffff)<<8);
+		src_data |= ((space->read_dword((src & 0x07fffffc)+4) & 0xff000000) >> 24);
+		src_data >>= (*src_shift)*16;
+	}
+	else
+		src_data = space->read_dword(src & 0x07fffffc) >> (*src_shift)*16;
+
+	space->write_word(dst,src_data);
+
+	*src_shift ^= 1;
+}
+
 static void scu_dma_direct(address_space *space, UINT8 dma_ch)
 {
 	saturn_state *state = space->machine().driver_data<saturn_state>();
 	UINT32 tmp_src,tmp_dst,tmp_size;
-	UINT8 single_dma_step;
 	UINT8 cd_transfer_flag;
 
-	if(state->m_scu.src_add[dma_ch] == 0 || state->m_scu.dst_add[dma_ch] != 4)
+	if(state->m_scu.src_add[dma_ch] == 0 || state->m_scu.dst_add[dma_ch] != 2)
 	{
 	if(LOG_SCU) printf("DMA lv %d transfer START\n"
 			             "Start %08x End %08x Size %04x\n",dma_ch,state->m_scu.src[dma_ch],state->m_scu.dst[dma_ch],state->m_scu.size[dma_ch]);
@@ -680,13 +699,12 @@ static void scu_dma_direct(address_space *space, UINT8 dma_ch)
 	/* max size */
 	if(state->m_scu.size[dma_ch] == 0) { state->m_scu.size[dma_ch] = (dma_ch == 0) ? 0x00100000 : 0x1000; }
 
-	/*set here the boundaries checks*/
-	/*...*/
-	if((state->m_scu.dst_add[dma_ch] != state->m_scu.src_add[dma_ch]) && (ABUS(dma_ch)))
+	/* Virtual Mahjong does transfers from A-Bus to work ram h with destination add value == 4, assume it's invalid */
+	if(state->m_scu.dst_add[dma_ch] >= 4 && (ABUS(dma_ch)))
 	{
-		logerror("A-Bus invalid transfer, sets to default\n");
+		printf("A-Bus invalid transfer, sets to default\n");
 		scu_add_tmp = (state->m_scu.dst_add[dma_ch]*0x100) | (state->m_scu.src_add[dma_ch]);
-		state->m_scu.dst_add[dma_ch] = state->m_scu.src_add[dma_ch] = 4;
+		state->m_scu.dst_add[dma_ch] = 2;
 		scu_add_tmp |= 0x80000000;
 	}
 
@@ -696,69 +714,39 @@ static void scu_dma_direct(address_space *space, UINT8 dma_ch)
 	if(!(DRUP(dma_ch))) tmp_src = state->m_scu.src[dma_ch];
 	if(!(DWUP(dma_ch))) tmp_dst = state->m_scu.dst[dma_ch];
 
-	single_dma_step = state->m_scu.dst_add[dma_ch];
 	cd_transfer_flag = state->m_scu.src_add[dma_ch] == 0 && state->m_scu.src[dma_ch] == 0x05818000;
 
-	if(single_dma_step > 4 && !cd_transfer_flag)
-		single_dma_step = 4;
-
-	if(single_dma_step == 2)
-		popmessage("Single DMA step == 2??? Contact MAMEdev");
-
-	/* Advanced World War screen */
-	if(state->m_scu.dst_add[dma_ch] == 0x40)
+	/* Many games directly accesses CD-ROM register 0x05818000, it must be a dword access with current implementation otherwise it won't work */
+	if(cd_transfer_flag)
 	{
-		for (; state->m_scu.size[dma_ch] > 0; state->m_scu.size[dma_ch]-=2)
+		int i;
+		state->m_scu.dst_add[dma_ch] <<= 1;
+
+		for (i = 0; i < state->m_scu.size[dma_ch];i+=state->m_scu.dst_add[dma_ch])
 		{
-			space->write_word(state->m_scu.dst[dma_ch],     space->read_word(state->m_scu.src[dma_ch]  ));
-			state->m_scu.dst[dma_ch]+=0x20;
-			state->m_scu.src[dma_ch]+=2;
+			space->write_dword(state->m_scu.dst[dma_ch],space->read_dword(state->m_scu.src[dma_ch]));
+			if(state->m_scu.dst_add[dma_ch] == 8)
+				space->write_dword(state->m_scu.dst[dma_ch]+4,space->read_dword(state->m_scu.src[dma_ch]));
+
+			state->m_scu.src[dma_ch]+=state->m_scu.src_add[dma_ch];
+			state->m_scu.dst[dma_ch]+=state->m_scu.dst_add[dma_ch];
 		}
 	}
 	else
 	{
-	for (; state->m_scu.size[dma_ch] > 0; state->m_scu.size[dma_ch]-=single_dma_step)
-	{
-		/* Many games directly accesses CD-ROM register 0x05818000, it must be a dword access with current implementation otherwise it won't work */
-		if(cd_transfer_flag)
-		{
-			space->write_dword(state->m_scu.dst[dma_ch],  space->read_dword(state->m_scu.src[dma_ch]  ));
-			if(state->m_scu.dst_add[dma_ch] == 8)
-				space->write_dword(state->m_scu.dst[dma_ch]+4,space->read_dword(state->m_scu.src[dma_ch]  ));
-		}
-		else if(state->m_scu.dst_add[dma_ch] == 2)
-			space->write_word(state->m_scu.dst[dma_ch],space->read_word(state->m_scu.src[dma_ch]));
-		else if(state->m_scu.dst_add[dma_ch] == 0x40)
-		{
-			space->write_word(state->m_scu.dst[dma_ch],     space->read_word(state->m_scu.src[dma_ch]  ));
-		}
-		else if(state->m_scu.dst_add[dma_ch] == 8)
-		{
-			/* TRUSTED, Battle Garegga and Batsugun graphics */
-			space->write_word(state->m_scu.dst[dma_ch],  space->read_word(state->m_scu.src[dma_ch]  ));
-			space->write_word(state->m_scu.dst[dma_ch]+4,space->read_word(state->m_scu.src[dma_ch]+2));
-		}
-		else if(state->m_scu.src[dma_ch] & 1) // odd address access? Road Blaster uses this for work-ram to color ram transfers ...
-		{
-			UINT16 src_data;
+		int i;
+		UINT8  src_shift;
 
-			src_data = ((space->read_word(state->m_scu.src[dma_ch]-1) & 0xff) << 8);
-			src_data|= ((space->read_word(state->m_scu.src[dma_ch]+1) & 0xff00) >> 8);
-			space->write_word(state->m_scu.dst[dma_ch],  src_data);
+		src_shift = ((state->m_scu.src[dma_ch] & 2) >> 1) ^ 1;
 
-			src_data = ((space->read_word(state->m_scu.src[dma_ch]+1) & 0xff) << 8);
-			src_data|= ((space->read_word(state->m_scu.src[dma_ch]+3) & 0xff00) >> 8);
-			space->write_word(state->m_scu.dst[dma_ch]+2,src_data);
-		}
-		else
+		for (i = 0; i < state->m_scu.size[dma_ch];i+=2)
 		{
-			space->write_word(state->m_scu.dst[dma_ch],  space->read_word(state->m_scu.src[dma_ch]  ));
-			space->write_word(state->m_scu.dst[dma_ch]+2,space->read_word(state->m_scu.src[dma_ch]+2));
-		}
+			scu_single_transfer(space,state->m_scu.src[dma_ch],state->m_scu.dst[dma_ch],&src_shift);
 
-		state->m_scu.dst[dma_ch]+=state->m_scu.dst_add[dma_ch];
-		state->m_scu.src[dma_ch]+=state->m_scu.src_add[dma_ch];
-	}
+			if(src_shift)
+				state->m_scu.src[dma_ch]+=state->m_scu.src_add[dma_ch];
+			state->m_scu.dst[dma_ch]+=state->m_scu.dst_add[dma_ch];
+		}
 	}
 
 	state->m_scu.size[dma_ch] = tmp_size;
@@ -809,7 +797,7 @@ static void scu_dma_indirect(address_space *space,UINT8 dma_ch)
 		if(indirect_src & 0x80000000)
 			job_done = 1;
 
-		if(state->m_scu.src_add[dma_ch] == 0 || state->m_scu.dst_add[dma_ch] != 4)
+		if(state->m_scu.src_add[dma_ch] == 0 || state->m_scu.dst_add[dma_ch] != 2)
 		{
 			if(LOG_SCU) printf("DMA lv %d indirect mode transfer START\n"
 					           "Index %08x Start %08x End %08x Size %04x\n",dma_ch,tmp_src,indirect_src,indirect_dst,indirect_size);
@@ -822,22 +810,20 @@ static void scu_dma_indirect(address_space *space,UINT8 dma_ch)
 
 		if(indirect_size == 0) { indirect_size = (dma_ch == 0) ? 0x00100000 : 0x2000; }
 
-		for (; indirect_size > 0; indirect_size-=state->m_scu.dst_add[dma_ch])
 		{
-			if(state->m_scu.dst_add[dma_ch] == 2)
-				space->write_word(indirect_dst,space->read_word(indirect_src));
-			else
+			int i;
+			UINT8  src_shift;
+
+			src_shift = ((indirect_src & 2) >> 1) ^ 1;
+
+			for (i = 0; i < indirect_size;i+=2)
 			{
-				/* some games, eg columns97 are a bit weird, I'm not sure this is correct
-                  they start a dma on a 2 byte boundary in 4 byte add mode, using the dword reads we
-                  can't access 2 byte boundaries, and the end of the sprite list never gets marked,
-                  the length of the transfer is also set to a 2 byte boundary, maybe the add values
-                  should be different, I don't know */
-				space->write_word(indirect_dst,space->read_word(indirect_src));
-				space->write_word(indirect_dst+2,space->read_word(indirect_src+2));
+				scu_single_transfer(space,indirect_src,indirect_dst,&src_shift);
+
+				if(src_shift)
+					indirect_src+=state->m_scu.src_add[dma_ch];
+				indirect_dst+=state->m_scu.dst_add[dma_ch];
 			}
-			indirect_dst+=state->m_scu.dst_add[dma_ch];
-			indirect_src+=state->m_scu.src_add[dma_ch];
 		}
 
 		//if(DRUP(0))   space->write_dword(tmp_src+8,state->m_scu.src[0]|job_done ? 0x80000000 : 0);
diff --git a/src/mame/includes/stv.h b/src/mame/includes/stv.h
index 3a7cfc3ac1a..f3abffe7777 100644
--- a/src/mame/includes/stv.h
+++ b/src/mame/includes/stv.h
@@ -27,7 +27,7 @@ public:
 		UINT32    dst[3];		/* Destination DMA lv n address*/
 		UINT32    src_add[3];	/* Source Addition for DMA lv n*/
 		UINT32    dst_add[3];	/* Destination Addition for DMA lv n*/
-		INT32     size[3];		/* Transfer DMA size lv n*/
+		UINT32    size[3];		/* Transfer DMA size lv n*/
 		UINT32    index[3];
 		int       start_factor[3];
 		UINT8     enable_mask[3];
diff --git a/src/mame/machine/smpc.c b/src/mame/machine/smpc.c
index e1a9a977f0f..2ce31520d3c 100644
--- a/src/mame/machine/smpc.c
+++ b/src/mame/machine/smpc.c
@@ -153,7 +153,7 @@ TODO:
 #include "machine/eeprom.h"
 
 #define LOG_SMPC 0
-#define LOG_PAD_CMD 0
+#define LOG_PAD_CMD 1
 
 READ8_HANDLER( stv_SMPC_r )
 {
@@ -324,12 +324,12 @@ static TIMER_CALLBACK( intback_peripheral )
 	static const char *const padnames[] = { "JOY1", "JOY2" };
 
 	/* doesn't work? */
-	//pad_num = state->m_smpc.intback_stage - 1;
+	pad_num = state->m_smpc.intback_stage - 1;
 
 	if(LOG_PAD_CMD) printf("%d\n",pad_num);
 
 //  if (LOG_SMPC) logerror("SMPC: providing PAD data for intback, pad %d\n", intback_stage-2);
-	for(pad_num=0;pad_num<2;pad_num++)
+	//for(pad_num=0;pad_num<2;pad_num++)
 	{
 		pad = input_port_read(machine, padnames[pad_num]);
 		state->m_smpc_ram[0x21+pad_num*8] = 0xf1;	// no tap, direct connect
@@ -417,6 +417,7 @@ static TIMER_CALLBACK( saturn_smpc_intback )
 		state->m_smpc.intback_stage = (state->m_smpc_ram[3] & 8) >> 3; // first peripheral
 		state->m_smpc.smpcSR = 0x40 | state->m_smpc.intback_stage << 5;
 		state->m_smpc.pmode = state->m_smpc_ram[1]>>4;
+		machine.scheduler().timer_set(attotime::from_usec(15000), FUNC(intback_peripheral),0);
 
 		if(!(state->m_scu.ism & IRQ_SMPC))
 			device_set_input_line_and_vector(state->m_maincpu, 8, HOLD_LINE, 0x47);
diff --git a/src/mame/video/stvvdp2.c b/src/mame/video/stvvdp2.c
index 135b6cabd67..3f838d18a98 100644
--- a/src/mame/video/stvvdp2.c
+++ b/src/mame/video/stvvdp2.c
@@ -4337,9 +4337,9 @@ static void stv_vdp2_check_tilemap(running_machine &machine, bitmap_t *bitmap, c
 			popmessage("Special Color Calculation enable %04x, contact MAMEdev",STV_VDP2_SFCCMD);
 
 		/* Cleopatra Fortune Transparent Shadow */
+		/* Pretty Fighter X Back & Transparent Shadow*/
 		//if(STV_VDP2_SDCTL & 0x0120)
-		if(STV_VDP2_SDCTL & 0x0020)
-			popmessage("%s shadow select bit enabled, contact MAMEdev",STV_VDP2_SDCTL & 0x100 ? "Transparent" : "Back");
+		//	popmessage("%s shadow select bit enabled, contact MAMEdev",STV_VDP2_SDCTL & 0x100 ? "Transparent" : "Back");
 
 		/* Langrisser III bit 3 normal, bit 1 during battle field */
 		if(STV_VDP2_SFSEL & ~0xa)