8 channel LPCM over HDMI, It's Multi Amp for Everyone - Page 9 - diyAudio
Go Back   Home > Forums > Source & Line > Digital Line Level

Digital Line Level DACs, Digital Crossovers, Equalizers, etc.

Please consider donating to help us continue to serve you.

Ads on/off / Custom Title / More PMs / More album space / Advanced printing & mass image saving
Reply
 
Thread Tools Search this Thread
Old 24th August 2008, 04:11 AM   #81
diyAudio Member
 
Join Date: Nov 2006
nuhi
http://www.microsoft.com/express/2005/

fb,
9500GT CUDA1.1 is float in / float out.
If you use GTX280, it can be double in / double out. but GTX280 is too noisy, (yes I have in another PC) to use in Audio PC.
Now I moved to CUDA 2.0.

TAPS: 8192 is not a limit, it can be 16384 without change of logic.
__device__ __constant__ float coeff_kernel
maximum size is 65536 bytes = 16384 float.
If you need more TAPS, coeff_kernel have to be placed in global.

I think 65536 TAPS is pretty bellyful? it has 0.67Hz filter pitch and 1.48second delay at 44100 sampling.
but "how long FIR taps can CUDA process?" is interesting challenge
I should build another VS2005 project for GTX280 later.

If you are in US I thought to send you GeForce8400!
but postage to AU is $30, equal to card itself.
  Reply With Quote
Old 24th August 2008, 06:52 AM   #82
diyAudio Member
 
Join Date: Nov 2006
please let me use here as memo, I was missimplementing delay/taps.

(1) TAP length and delay
Current, new incoming buffer is used for FIR processing.
[NNNNNNNN] requires = TAPS length

New: ex, 1024 samples frame incoming and kept 15360 samples
[PPPPPPPPPPPPPPPN], delay is 1024 sample.
Prev 1024*15 samples + New 1024 samples are used for 16384 length FIR. [PPPPPPPPPPPPPPPN] is stored and manipulated in host memory. it will work as delay line ( can contain channel delay logic).

(2) FIR Calling Sequence
Current, kernel called 8 times per frame. and taps limited 16384 by constants memory size.
for (way 0,1,2,3)
for (ch L, R)
copy FIR Coeff to constants
do FIR<<kernel>>

New, kernel call once per frame, no taps length limit
Copy Coeff to Global (once )
While()
wait Event
Copy [PPPPPPPPPPPPPPN] to device
Execute <<kernel>>(global, global)
Copy result to Host
Loopend

(3) Input stream status
by paInputUnderflow, paAbort, stream stop.
add input stream check, pass 00000000 to GPU.
  Reply With Quote
Old 24th August 2008, 01:20 PM   #83
fb is offline fb  Australia
diyAudio Member
 
Join Date: Apr 2005
Location: Brisbane, Aus
Quote:
Originally posted by KOON3876

fb,
9500GT CUDA1.1 is float in / float out.
If you use GTX280, it can be double in / double out. but GTX280 is too noisy, (yes I have in another PC) to use in Audio PC.
Now I moved to CUDA 2.0.

If you are in US I thought to send you GeForce8400!
but postage to AU is $30, equal to card itself.

Thanks I'll buy a card, just wanting to be somewhat future proof I'm not good at understanding the maths side of it.... can 9500gt do 24bit?
  Reply With Quote
Old 24th August 2008, 01:25 PM   #84
nuhi is offline nuhi  Croatia
diyAudio Member
 
Join Date: May 2005
Location: Croatia
fb, aha so just change compiler path to VS8 and use VS9, good.

Koon, what kind of answer is that, did you even read what I asked...this is the second time you give me a half-a$$ed answer and then I need to cool down, yeah right.
  Reply With Quote
Old 24th August 2008, 06:09 PM   #85
diyAudio Member
 
Join Date: Nov 2006
fb, accuracy of float, below shows what you want?

f2 = (float)X1 / (float)( 65536 * 128 );
printf("Int %d\n", X1);
printf("float %2.16f\n", f2);
printf("converted back to int %d\n", (long)(f2*(65536 * 128)));


Int 1
float 0.0000001192092896
converted back to int 1
Int 16
float 0.0000019073486328
converted back to int 16
Int 151
float 0.0000180006027222
converted back to int 151
Int 1366
float 0.0001628398895264
converted back to int 1366
Int 12301
float 0.0014663934707642
converted back to int 12301
Int 110716
float 0.0131983757019043
converted back to int 110716
Int 996451
float 0.1187862157821655
converted back to int 996451


nuhi, I just post URL how to get 2005. That's simplest way to Compile CUDA. why you are so complaint? This is DIYaudio.com.
  Reply With Quote
Old 24th August 2008, 06:30 PM   #86
nuhi is offline nuhi  Croatia
diyAudio Member
 
Join Date: May 2005
Location: Croatia
Koon, you got to be kidding me, you really don't understand what I'm asking. Fine, I won't post in this topic any more.
  Reply With Quote
Old 24th August 2008, 10:07 PM   #87
diyAudio Member
 
Join Date: Nov 2006
???
He mentioned he can't compile with VS2008 and he knows what he wants to do, and I posted free VS2005 express URL.

Or does he want ME to build another version as he like??
  Reply With Quote
Old 25th August 2008, 04:59 AM   #88
diyAudio Member
 
Join Date: Nov 2006
(1) 131072 TAPS test
It looks like GTX280 can process 131072 taps FIR in realtime.
================
Using device 0: GeForce GTX 280
[RIFF] (28376916 bytes)
[WAVEfmt ] (16 bytes)
[data] (28376880 bytes)
total samples = 7094220, music length = 160 sec.
Started 970ms
Calculating 14 / 14
done FIR processing.
Ended 113308ms
elapsed 112 sec.
================
required performance
131072(tap)*44100(sample/sec)*2(channel)*4(way)*2(op) = 92GFlops,
achieved performance
92GFlops * 160/112 = 131GFlops
... Still I'm not using full power of GPU. I need to study streaming or concurrent operation. (now GPU is idle when copying something from/to)

(2) Errors I see
connect / disconnect SPDIF can make error below. or I have to run thousands seconds to see.

type 1: sudden stop of stream, patestCallback is not called in cyclic.
I added WaitForSingleObject(patestEvent, 740ms); to find this case. I have to re-initialize buffer and stream.
type 2: PaStreamCallBackFlag 0x0001 paInputUnderFlow
input is not ready when callback called?

Still I'm doing long-run test now, and I see type(2) error at 3,541sec, type(1) error at 5,309sec.
===============================
anyway there are headroom for performance, errors to be resolved.
  Reply With Quote
Old 25th August 2008, 06:04 AM   #89
phofman is offline phofman  Czech Republic
diyAudio Member
 
Join Date: Apr 2005
Location: Pilsen
I am just remotely watching your thread, amazing work. Your input (spdif-in) and output (hdmi) clocks are likely unsynchronized. I am afraid in the long run that will be causing buffer under/overruns too. Though when taking into account the delay introduced by the long filter, a few thousand samples long buffer could postpone the mismatch for a few minutes. My CD player and sound card clocks differred one sample a second on 44.1kHz.
  Reply With Quote
Old 25th August 2008, 03:17 PM   #90
diyAudio Member
 
Join Date: Nov 2006
Quote:
Originally posted by phofman
I am afraid in the long run that will be causing buffer under/overruns too.
Thank you, also I'm afraid of unsync now.
44102 / 44100 will consume all buffer length 8192, in 4096 seconds. ... it looks like my case.

Now I made "Dummy In" "Dummy Out" mode test program.
Logics are same, but Dummy In works only with OUT stream. input is always 0000 from dummy. Dummy Out works only with IN stream, output is to dummy.
I started "Dummy In" version now - will run for 8 hours
  Reply With Quote

Reply


Hide this!Advertise here!
Thread Tools Search this Thread
Search this Thread:

Advanced Search

Posting Rules
You may not post new threads
You may not post replies
You may not post attachments
You may not edit your posts

BB code is On
Smilies are On
[IMG] code is On
HTML code is Off
Trackbacks are Off
Pingbacks are Off
Refbacks are Off


Similar Threads
Thread Thread Starter Forum Replies Last Post
Creating 5.1 Multi Channel sound from 2 channel jenkrich Everything Else 5 1st September 2009 08:29 AM
Multi purpose multi channel gainclone PCB: Team project - group buy rick57 Chip Amps 14 29th March 2003 06:54 PM
multi-channel AMP? JAZZ2250 Solid State 1 15th September 2002 10:10 AM


New To Site? Need Help?

All times are GMT. The time now is 04:54 AM.


vBulletin Optimisation provided by vB Optimise (Pro) - vBulletin Mods & Addons Copyright © 2014 DragonByte Technologies Ltd.
Copyright ©1999-2014 diyAudio

Content Relevant URLs by vBSEO 3.3.2